fdo-mirrors/mesa
1
0
Fork 0
mirror of https://gitlab.freedesktop.org/mesa/mesa.git synced 2025-12-23 02:30:12 +01:00
Code Issues Projects Releases Packages Wiki Activity Actions 1

spirv: Rework handling of spec constant workgroup size built-ins

Browse source 
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 <caio.oliveira@intel.com>
This commit is contained in:
Jason Ekstrand 2019-01-23 12:49:15 -06:00 • committed by Jason Ekstrand
parent 9b37e93e42
commit 7d862ef530
2 changed files with 14 additions and 4 deletions
Show stats Download patch file Download diff file Expand all files Collapse all files

17
src/compiler/spirv/spirv_to_nir.c
View file

@ -1564,10 +1564,7 @@ handle_workgroup_size_decoration_cb(struct vtn_builder *b,
return; return;
vtn_assert(val->type->type == glsl_vector_type(GLSL_TYPE_UINT, 3)); vtn_assert(val->type->type == glsl_vector_type(GLSL_TYPE_UINT, 3));
b->workgroup_size_builtin = val;
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];
} }
static void 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, words = vtn_foreach_instruction(b, words, word_end,
vtn_handle_variable_or_type_instruction); 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 */ /* Set types on all vtn_values */
vtn_foreach_instruction(b, words, word_end, vtn_set_instruction_result_type); vtn_foreach_instruction(b, words, word_end, vtn_set_instruction_result_type);

1
src/compiler/spirv/vtn_private.h
View file

@ -600,6 +600,7 @@ struct vtn_builder {
gl_shader_stage entry_point_stage; gl_shader_stage entry_point_stage;
const char *entry_point_name; const char *entry_point_name;
struct vtn_value *entry_point; struct vtn_value *entry_point;
struct vtn_value *workgroup_size_builtin;
bool origin_upper_left; bool origin_upper_left;
bool pixel_center_integer; bool pixel_center_integer;
bool variable_pointers; bool variable_pointers;