AuroraMiddleware/SPIRV-Cross
1
0
Fork 0
Code Issues Pull Requests Packages Projects Releases Wiki Activity

Reflect: Deal with workgroup size being specialization constants.

Browse Source
This commit is contained in:
Hans-Kristian Arntzen 2019-10-04 10:37:47 +02:00
parent 1a20737af7
commit 43e89bd269
13 changed files with 144 additions and 4 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

5
reference/shaders-reflection/asm/op-source-glsl-ssbo-1.asm.comp.json
View File

@ -7,6 +7,11 @@
1,
1,
1
],
"workgroup_size_is_spec_constant_id" : [
false,
false,
false
]
}
],

5
reference/shaders-reflection/asm/op-source-glsl-ssbo-2.asm.comp.json
View File

@ -7,6 +7,11 @@
1,
1,
1
],
"workgroup_size_is_spec_constant_id" : [
false,
false,
false
]
}
],

5
reference/shaders-reflection/asm/op-source-hlsl-uav-1.asm.comp.json
View File

@ -7,6 +7,11 @@
1,
1,
1
],
"workgroup_size_is_spec_constant_id" : [
false,
false,
false
]
}
],

5
reference/shaders-reflection/asm/op-source-hlsl-uav-2.asm.comp.json
View File

@ -7,6 +7,11 @@
1,
1,
1
],
"workgroup_size_is_spec_constant_id" : [
false,
false,
false
]
}
],

5
reference/shaders-reflection/asm/op-source-none-ssbo-1.asm.comp.json
View File

@ -7,6 +7,11 @@
1,
1,
1
],
"workgroup_size_is_spec_constant_id" : [
false,
false,
false
]
}
],

5
reference/shaders-reflection/asm/op-source-none-ssbo-2.asm.comp.json
View File

@ -7,6 +7,11 @@
1,
1,
1
],
"workgroup_size_is_spec_constant_id" : [
false,
false,
false
]
}
],

5
reference/shaders-reflection/asm/op-source-none-uav-1.asm.comp.json
View File

@ -7,6 +7,11 @@
1,
1,
1
],
"workgroup_size_is_spec_constant_id" : [
false,
false,
false
]
}
],

5
reference/shaders-reflection/asm/op-source-none-uav-2.asm.comp.json
View File

@ -7,6 +7,11 @@
1,
1,
1
],
"workgroup_size_is_spec_constant_id" : [
false,
false,
false
]
}
],

5
reference/shaders-reflection/comp/struct-layout.comp.json
View File

@ -7,6 +7,11 @@
1,
1,
1
],
"workgroup_size_is_spec_constant_id" : [
false,
false,
false
]
}
],

5
reference/shaders-reflection/comp/struct-packing.comp.json
View File

@ -7,6 +7,11 @@
1,
1,
1
],
"workgroup_size_is_spec_constant_id" : [
false,
false,
false
]
}
],

56
reference/shaders-reflection/comp/workgroup-size-spec-constant.comp.json Normal file
View File

@ -0,0 +1,56 @@
{
"entryPoints" : [
{
"name" : "main",
"mode" : "comp",
"workgroup_size" : [
10,
40,
60
],
"workgroup_size_is_spec_constant_id" : [
true,
true,
true
]
}
],
"types" : {
"_8" : {
"name" : "SSBO",
"members" : [
{
"name" : "v",
"type" : "vec4",
"offset" : 0
}
]
}
},
"ssbos" : [
{
"type" : "_8",
"name" : "SSBO",
"block_size" : 16,
"set" : 0,
"binding" : 0
}
],
"specialization_constants" : [
{
"id" : 10,
"type" : "uint",
"default_value" : 1
},
{
"id" : 40,
"type" : "uint",
"default_value" : 1
},
{
"id" : 60,
"type" : "uint",
"default_value" : 1
}
]
}

13
shaders-reflection/comp/workgroup-size-spec-constant.comp Normal file
View File

@ -0,0 +1,13 @@
#version 450
layout(local_size_x_id = 10, local_size_y_id = 40, local_size_z_id = 60) in;
layout(std430, set = 0, binding = 0) buffer SSBO
{
vec4 v;
};
void main()
{
v = vec4(10.0);
}

29
spirv_reflect.cpp
View File

@ -61,6 +61,7 @@ public:
void end_json_array();
void emit_json_array_value(const std::string &value);
void emit_json_array_value(uint32_t value);
void emit_json_array_value(bool value);
std::string str() const
{
@ -158,6 +159,16 @@ void Stream::emit_json_array_value(uint32_t value)
stack.top().second = true;
}
void Stream::emit_json_array_value(bool value)
{
if (stack.empty() || stack.top().first != Type::Array)
SPIRV_CROSS_THROW("Invalid JSON state");
if (stack.top().second)
statement_inner(",\n");
statement_no_return(value ? "true" : "false");
stack.top().second = true;
}
void Stream::begin_json_object()
{
if (!stack.empty() && stack.top().second)
@ -424,13 +435,23 @@ void CompilerReflection::emit_entry_points()
json_stream->begin_json_object();
json_stream->emit_json_key_value("name", e.name);
json_stream->emit_json_key_value("mode", execution_model_to_str(e.execution_model));
if(e.execution_model == ExecutionModelGLCompute)
if (e.execution_model == ExecutionModelGLCompute)
{
const auto &spv_entry = get_entry_point(e.name, e.execution_model);
SpecializationConstant spec_x, spec_y, spec_z;
get_work_group_size_specialization_constants(spec_x, spec_y, spec_z);
json_stream->emit_json_key_array("workgroup_size");
json_stream->emit_json_array_value(spv_entry.workgroup_size.x);
json_stream->emit_json_array_value(spv_entry.workgroup_size.y);
json_stream->emit_json_array_value(spv_entry.workgroup_size.z);
json_stream->emit_json_array_value(spec_x.id != ID(0) ? spec_x.constant_id : spv_entry.workgroup_size.x);
json_stream->emit_json_array_value(spec_y.id != ID(0) ? spec_y.constant_id : spv_entry.workgroup_size.y);
json_stream->emit_json_array_value(spec_z.id != ID(0) ? spec_z.constant_id : spv_entry.workgroup_size.z);
json_stream->end_json_array();
json_stream->emit_json_key_array("workgroup_size_is_spec_constant_id");
json_stream->emit_json_array_value(spec_x.id != ID(0));
json_stream->emit_json_array_value(spec_y.id != ID(0));
json_stream->emit_json_array_value(spec_z.id != ID(0));
json_stream->end_json_array();
}
json_stream->end_json_object();