From 7d862ef53024b70458929c1665af43573b4d07ff Mon Sep 17 00:00:00 2001 From: Jason Ekstrand Date: Wed, 23 Jan 2019 12:49:15 -0600 Subject: [PATCH] spirv: Rework handling of spec constant workgroup size built-ins Instead of handling it as part of the handling of constant instructions, just stash the vtn_value when we see the decoration and handle it explicitly later. This will let us re-order handling of constant instructions without breaking the Vulkan SPIR-V requirement that decorating a specialization constant as the WorkgroupSize built-in overrides the workgroup size set as an execution mode. Reviewed-by: Caio Marcelo de Oliveira Filho --- src/compiler/spirv/spirv_to_nir.c | 17 +++++++++++++---- src/compiler/spirv/vtn_private.h | 1 + 2 files changed, 14 insertions(+), 4 deletions(-) diff --git a/src/compiler/spirv/spirv_to_nir.c b/src/compiler/spirv/spirv_to_nir.c index 1a26c51..41972ed 100644 --- a/src/compiler/spirv/spirv_to_nir.c +++ b/src/compiler/spirv/spirv_to_nir.c @@ -1564,10 +1564,7 @@ handle_workgroup_size_decoration_cb(struct vtn_builder *b, return; vtn_assert(val->type->type == glsl_vector_type(GLSL_TYPE_UINT, 3)); - - b->shader->info.cs.local_size[0] = val->constant->values[0].u32[0]; - b->shader->info.cs.local_size[1] = val->constant->values[0].u32[1]; - b->shader->info.cs.local_size[2] = val->constant->values[0].u32[2]; + b->workgroup_size_builtin = val; } static void @@ -4455,6 +4452,18 @@ spirv_to_nir(const uint32_t *words, size_t word_count, words = vtn_foreach_instruction(b, words, word_end, vtn_handle_variable_or_type_instruction); + if (b->workgroup_size_builtin) { + vtn_assert(b->workgroup_size_builtin->type->type == + glsl_vector_type(GLSL_TYPE_UINT, 3)); + + nir_const_value *const_size = + &b->workgroup_size_builtin->constant->values[0]; + + b->shader->info.cs.local_size[0] = const_size->u32[0]; + b->shader->info.cs.local_size[1] = const_size->u32[1]; + b->shader->info.cs.local_size[2] = const_size->u32[2]; + } + /* Set types on all vtn_values */ vtn_foreach_instruction(b, words, word_end, vtn_set_instruction_result_type); diff --git a/src/compiler/spirv/vtn_private.h b/src/compiler/spirv/vtn_private.h index 07e3311..6331303 100644 --- a/src/compiler/spirv/vtn_private.h +++ b/src/compiler/spirv/vtn_private.h @@ -600,6 +600,7 @@ struct vtn_builder { gl_shader_stage entry_point_stage; const char *entry_point_name; struct vtn_value *entry_point; + struct vtn_value *workgroup_size_builtin; bool origin_upper_left; bool pixel_center_integer; bool variable_pointers; -- 2.7.4