AlexIndustrial/mesa
1
0
Fork 0
Code Issues Pull Requests Actions Packages Projects Releases Wiki Activity

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
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;
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);

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

@@ -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;