NeroReflex/mesa
1
0
Fork 0
Code Issues Pull requests Projects Releases Packages Wiki Activity
mesa/src/compiler/spirv
History
Pierre Moreau 4a408ff7ea spirv: Ignore WorkgroupSize in non-compute stages 
If a SPIR-V module contains for example both a geometry and a compute
shader, when processing the geometry shader its vertices out, input
primitive and output primitive attributes would get overwritten by the
value of the WorkgroupSize.

```
; SPIR-V
; Version: 1.5
; Generator: Khronos; 17
; Bound: 12
; Schema: 0
               OpCapability Geometry
               OpCapability Shader
          %1 = OpExtInstImport "GLSL.std.450"
               OpMemoryModel Logical GLSL450
               OpEntryPoint Geometry %main "main"
               OpEntryPoint GLCompute %main_0 "main"
               OpExecutionMode %main InputPoints
               OpExecutionMode %main Invocations 1
               OpExecutionMode %main OutputTriangleStrip
               OpExecutionMode %main OutputVertices 4
               OpExecutionMode %main_0 LocalSize 1 1 1
               OpSource GLSL 460
               OpSource GLSL 460
               OpName %main "main"
               OpName %main_0 "main"
               OpModuleProcessed "Linked by SPIR-V Tools Linker"
               OpDecorate %gl_WorkGroupSize BuiltIn WorkgroupSize
       %void = OpTypeVoid
          %6 = OpTypeFunction %void
       %uint = OpTypeInt 32 0
     %v3uint = OpTypeVector %uint 3
     %uint_1 = OpConstant %uint 1
%gl_WorkGroupSize = OpConstantComposite %v3uint %uint_1 %uint_1 %uint_1
       %main = OpFunction %void None %6
         %10 = OpLabel
               OpReturn
               OpFunctionEnd
     %main_0 = OpFunction %void None %6
         %11 = OpLabel
               OpReturn
               OpFunctionEnd
```

Running spirv_to_nir on the SPIR-V sample above and for the geometry
entry point would say that (among others):

* vertices out: 1
* input primitive: LINES
* output primitive: LINES

By removing any reference to `%gl_WorkGroupSize`, the output would
change to (among others):

* vertices out: 4
* input primitive: POINTS
* output primitive: TRIANGLE_STRIP

Fixes: 7d862ef530 ("spirv: Rework handling of spec constant workgroup size built-ins")

v2:
* Move the check from inside `handle_workgroup_size_decoration_cb()` to
  its caller (Caio Marcelo de Oliveira Filho )
* Add an assert on the shader stage before using
  `workgroup_size_builtin` (Caio Marcelo de Oliveira Filho )

Reviewed-by: Caio Marcelo de Oliveira Filho <caio.oliveira@intel.com>
Signed-off-by: Pierre Moreau <dev@pmoreau.org>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/9418>
2021-03-11 20:30:38 +00:00
..
tests spirv: Initialize spirv_test member shader. 2020-09-09 22:24:09 +00:00
gl_spirv.c spirv: workaround setjmp/longjmp crash on MinGW 2020-11-18 13:53:36 +00:00
GLSL.ext.AMD.h
GLSL.std.450.h
meson.build spirv: add and use a generator id enum 2020-10-12 11:07:38 +00:00
nir_load_libclc.c nir_load_libclc: Mark libclc shader as internal 2020-11-18 04:05:37 +00:00
nir_lower_libclc.c spirv: Move nir_lower_libclc to src/compiler/spirv 2020-10-07 21:52:04 +00:00
nir_spirv.h spirv: Implement OpArrayLength for OpenGL 2020-12-18 17:13:46 +00:00
OpenCL.std.h
spir-v.xml spirv: Update headers and metadata from latest Khronos commit 2021-01-27 22:20:53 +00:00
spirv.core.grammar.json spirv: Update headers and metadata from latest Khronos commit 2021-01-27 22:20:53 +00:00
spirv.h spirv: Update headers and metadata from latest Khronos commit 2021-01-27 22:20:53 +00:00
spirv2nir.c spirv2nir: Add --opengl (-g) argument for OpenGL SPIR-V 2020-12-22 09:34:00 -08:00
spirv_info.h spirv: vtn_fail with a nice message on unsupported rounding modes 2020-09-25 01:43:28 +00:00
spirv_info_c.py spirv: vtn_fail with a nice message on unsupported rounding modes 2020-09-25 01:43:28 +00:00
spirv_to_nir.c spirv: Ignore WorkgroupSize in non-compute stages 2021-03-11 20:30:38 +00:00
vtn_alu.c spirv: handle NoContraction in GLSL450 alu ops 2021-01-23 01:39:09 +00:00
vtn_amd.c spirv: use intrinsic builders 2020-11-26 17:50:38 +00:00
vtn_cfg.c spirv: Delete the impl for prototype-only functions 2021-02-16 20:50:51 +00:00
vtn_gather_types_c.py spirv: Handle instruction aliases in vtn_gather_types 2020-04-24 05:56:05 +00:00
vtn_generator_ids_h.py android: fix SPIR-V -> NIR build 2020-10-12 22:26:05 +00:00
vtn_glsl450.c spirv: handle NoContraction in GLSL450 alu ops 2021-01-23 01:39:09 +00:00
vtn_opencl.c spirv: Store the nir_function in vtn_function 2021-02-16 20:50:51 +00:00
vtn_private.h spirv: Store the nir_function in vtn_function 2021-02-16 20:50:51 +00:00
vtn_subgroup.c spirv: Implement SpvCapabilitySubgroupShuffleINTEL from SPV_INTEL_subgroups 2020-11-04 20:24:48 +00:00
vtn_variables.c spirv: Update a couple of comments in variable handling 2021-03-08 20:23:28 +00:00