diff options
| author | Jason Ekstrand <jason.ekstrand@intel.com> | 2016-06-01 11:20:22 -0700 |
|---|---|---|
| committer | Jason Ekstrand <jason.ekstrand@intel.com> | 2016-06-03 19:29:28 -0700 |
| commit | 1f7b54ed299bea95f774e7d8baa181c11118b3fe (patch) | |
| tree | 05f28bf46f5dd9ce0ff4fde066cb0c4b23ba83f8 /src/compiler/spirv | |
| parent | b26cdd65e8e588ed7bdd7cca0b205579132e1261 (diff) | |
| download | external_mesa3d-1f7b54ed299bea95f774e7d8baa181c11118b3fe.zip external_mesa3d-1f7b54ed299bea95f774e7d8baa181c11118b3fe.tar.gz external_mesa3d-1f7b54ed299bea95f774e7d8baa181c11118b3fe.tar.bz2 | |
nir/spirv: Handle the WorkgroupSize builtin decoration
This fixes the 7 dEQP-VK.pipeline.spec_constant.compute.local_size.* tests
in the latest dev version of the Vulkan CTS.
Signed-off-by: Jason Ekstrand <jason@jlekstrand.net>
Reviewed-by: Kenneth Graunke <kenneth@whitecape.org>
Cc: "12.0" <mesa-stable@lists.freedesktop.org>
Diffstat (limited to 'src/compiler/spirv')
| -rw-r--r-- | src/compiler/spirv/spirv_to_nir.c | 22 |
1 files changed, 22 insertions, 0 deletions
diff --git a/src/compiler/spirv/spirv_to_nir.c b/src/compiler/spirv/spirv_to_nir.c index bb7aba4..cece645 100644 --- a/src/compiler/spirv/spirv_to_nir.c +++ b/src/compiler/spirv/spirv_to_nir.c @@ -932,6 +932,25 @@ get_specialization(struct vtn_builder *b, struct vtn_value *val, } static void +handle_workgroup_size_decoration_cb(struct vtn_builder *b, + struct vtn_value *val, + int member, + const struct vtn_decoration *dec, + void *data) +{ + assert(member == -1); + if (dec->decoration != SpvDecorationBuiltIn || + dec->literals[0] != SpvBuiltInWorkgroupSize) + return; + + assert(val->const_type == glsl_vector_type(GLSL_TYPE_UINT, 3)); + + b->shader->info.cs.local_size[0] = val->constant->value.u[0]; + b->shader->info.cs.local_size[1] = val->constant->value.u[1]; + b->shader->info.cs.local_size[2] = val->constant->value.u[2]; +} + +static void vtn_handle_constant(struct vtn_builder *b, SpvOp opcode, const uint32_t *w, unsigned count) { @@ -1151,6 +1170,9 @@ vtn_handle_constant(struct vtn_builder *b, SpvOp opcode, default: unreachable("Unhandled opcode"); } + + /* Now that we have the value, update the workgroup size if needed */ + vtn_foreach_decoration(b, val, handle_workgroup_size_decoration_cb, NULL); } static void |