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

MSL: Unroll initializations of CullDistance/ClipDistance control points.

Browse Source
This commit is contained in:
Hans-Kristian Arntzen 2021-04-14 12:08:43 +02:00
parent c9946296dd
commit b442500204
7 changed files with 36 additions and 16 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

4
reference/shaders-msl-no-opt/asm/masking/initializers-block.mask-point-size.msl2.asm.tesc
View File

@ -87,8 +87,8 @@ kernel void main0(uint gl_InvocationID [[thread_index_in_threadgroup]], uint gl_
device main0_out* gl_out = &spvOut[gl_PrimitiveID * 4];
gl_out[gl_InvocationID].C_v = _18[gl_InvocationID].v;
gl_out[gl_InvocationID].gl_Position = _33[gl_InvocationID].gl_Position;
gl_out[gl_InvocationID].gl_ClipDistance = _33[gl_InvocationID].gl_ClipDistance;
gl_out[gl_InvocationID].gl_CullDistance = _33[gl_InvocationID].gl_CullDistance;
gl_out[gl_InvocationID].gl_ClipDistance[0] = _33[gl_InvocationID].gl_ClipDistance[0];
gl_out[gl_InvocationID].gl_CullDistance[0] = _33[gl_InvocationID].gl_CullDistance[0];
gl_out_masked[gl_InvocationID] = _33[gl_InvocationID];
device main0_patchOut& patchOut = spvPatchOut[gl_PrimitiveID];
patchOut.P_v = float4(0.0);

4
reference/shaders-msl-no-opt/asm/masking/initializers-block.mask-position.msl2.asm.tesc
View File

@ -87,8 +87,8 @@ kernel void main0(uint gl_InvocationID [[thread_index_in_threadgroup]], uint gl_
device main0_out* gl_out = &spvOut[gl_PrimitiveID * 4];
gl_out[gl_InvocationID].C_v = _18[gl_InvocationID].v;
gl_out[gl_InvocationID].gl_PointSize = _33[gl_InvocationID].gl_PointSize;
gl_out[gl_InvocationID].gl_ClipDistance = _33[gl_InvocationID].gl_ClipDistance;
gl_out[gl_InvocationID].gl_CullDistance = _33[gl_InvocationID].gl_CullDistance;
gl_out[gl_InvocationID].gl_ClipDistance[0] = _33[gl_InvocationID].gl_ClipDistance[0];
gl_out[gl_InvocationID].gl_CullDistance[0] = _33[gl_InvocationID].gl_CullDistance[0];
gl_out_masked[gl_InvocationID] = _33[gl_InvocationID];
device main0_patchOut& patchOut = spvPatchOut[gl_PrimitiveID];
patchOut.P_v = float4(0.0);

4
reference/shaders-msl-no-opt/asm/masking/initializers.mask-location-0.msl2.asm.tesc
View File

@ -78,8 +78,8 @@ kernel void main0(uint gl_InvocationID [[thread_index_in_threadgroup]], uint gl_
foo[gl_InvocationID] = _15[gl_InvocationID];
gl_out[gl_InvocationID].gl_Position = _29[gl_InvocationID]._RESERVED_IDENTIFIER_FIXUP_gl_Position;
gl_out[gl_InvocationID].gl_PointSize = _29[gl_InvocationID]._RESERVED_IDENTIFIER_FIXUP_gl_PointSize;
gl_out[gl_InvocationID].gl_ClipDistance = _29[gl_InvocationID]._RESERVED_IDENTIFIER_FIXUP_gl_ClipDistance;
gl_out[gl_InvocationID].gl_CullDistance = _29[gl_InvocationID]._RESERVED_IDENTIFIER_FIXUP_gl_CullDistance;
gl_out[gl_InvocationID].gl_ClipDistance[0] = _29[gl_InvocationID]._RESERVED_IDENTIFIER_FIXUP_gl_ClipDistance[0];
gl_out[gl_InvocationID].gl_CullDistance[0] = _29[gl_InvocationID]._RESERVED_IDENTIFIER_FIXUP_gl_CullDistance[0];
device main0_patchOut& patchOut = spvPatchOut[gl_PrimitiveID];
patchOut.foo_patch = float4(0.0);
foo[gl_InvocationID] = float4(1.0);

4
reference/shaders-msl-no-opt/asm/masking/initializers.mask-location-1.asm.tesc
View File

@ -77,8 +77,8 @@ kernel void main0(uint gl_InvocationID [[thread_index_in_threadgroup]], uint gl_
gl_out[gl_InvocationID].foo = _15[gl_InvocationID];
gl_out[gl_InvocationID].gl_Position = _29[gl_InvocationID]._RESERVED_IDENTIFIER_FIXUP_gl_Position;
gl_out[gl_InvocationID].gl_PointSize = _29[gl_InvocationID]._RESERVED_IDENTIFIER_FIXUP_gl_PointSize;
gl_out[gl_InvocationID].gl_ClipDistance = _29[gl_InvocationID]._RESERVED_IDENTIFIER_FIXUP_gl_ClipDistance;
gl_out[gl_InvocationID].gl_CullDistance = _29[gl_InvocationID]._RESERVED_IDENTIFIER_FIXUP_gl_CullDistance;
gl_out[gl_InvocationID].gl_ClipDistance[0] = _29[gl_InvocationID]._RESERVED_IDENTIFIER_FIXUP_gl_ClipDistance[0];
gl_out[gl_InvocationID].gl_CullDistance[0] = _29[gl_InvocationID]._RESERVED_IDENTIFIER_FIXUP_gl_CullDistance[0];
device main0_patchOut& patchOut = spvPatchOut[gl_PrimitiveID];
foo_patch = float4(0.0);
gl_out[gl_InvocationID].foo = float4(1.0);

4
reference/shaders-msl-no-opt/asm/masking/initializers.mask-point-size.msl2.asm.tesc
View File

@ -77,8 +77,8 @@ kernel void main0(uint gl_InvocationID [[thread_index_in_threadgroup]], uint gl_
device main0_out* gl_out = &spvOut[gl_PrimitiveID * 4];
gl_out[gl_InvocationID].foo = _15[gl_InvocationID];
gl_out[gl_InvocationID].gl_Position = _29[gl_InvocationID].gl_Position;
gl_out[gl_InvocationID].gl_ClipDistance = _29[gl_InvocationID].gl_ClipDistance;
gl_out[gl_InvocationID].gl_CullDistance = _29[gl_InvocationID].gl_CullDistance;
gl_out[gl_InvocationID].gl_ClipDistance[0] = _29[gl_InvocationID].gl_ClipDistance[0];
gl_out[gl_InvocationID].gl_CullDistance[0] = _29[gl_InvocationID].gl_CullDistance[0];
gl_out_masked[gl_InvocationID] = _29[gl_InvocationID];
device main0_patchOut& patchOut = spvPatchOut[gl_PrimitiveID];
patchOut.foo_patch = float4(0.0);

4
reference/shaders-msl-no-opt/asm/masking/initializers.mask-position.msl2.asm.tesc
View File

@ -77,8 +77,8 @@ kernel void main0(uint gl_InvocationID [[thread_index_in_threadgroup]], uint gl_
device main0_out* gl_out = &spvOut[gl_PrimitiveID * 4];
gl_out[gl_InvocationID].foo = _15[gl_InvocationID];
gl_out[gl_InvocationID].gl_PointSize = _29[gl_InvocationID].gl_PointSize;
gl_out[gl_InvocationID].gl_ClipDistance = _29[gl_InvocationID].gl_ClipDistance;
gl_out[gl_InvocationID].gl_CullDistance = _29[gl_InvocationID].gl_CullDistance;
gl_out[gl_InvocationID].gl_ClipDistance[0] = _29[gl_InvocationID].gl_ClipDistance[0];
gl_out[gl_InvocationID].gl_CullDistance[0] = _29[gl_InvocationID].gl_CullDistance[0];
gl_out_masked[gl_InvocationID] = _29[gl_InvocationID];
device main0_patchOut& patchOut = spvPatchOut[gl_PrimitiveID];
patchOut.foo_patch = float4(0.0);

28
spirv_msl.cpp
View File

@ -2770,13 +2770,33 @@ void CompilerMSL::add_plain_member_variable_to_interface_block(StorageClass stor
auto &type = this->get<SPIRType>(var.basetype);
uint32_t index = get_extended_member_decoration(var.self, mbr_idx, SPIRVCrossDecorationInterfaceMemberIndex);
auto &mbr_type = this->get<SPIRType>(type.member_types[mbr_idx]);
bool unroll_array = !mbr_type.array.empty() && is_builtin;
AccessChainMeta chain_meta;
auto constant_chain = access_chain_internal(var.initializer, &builtin_invocation_id_id, 1, 0, &chain_meta);
statement(to_expression(stage_out_ptr_var_id), "[",
builtin_to_glsl(BuiltInInvocationId, StorageClassInput), "].",
to_member_name(ib_type, index), " = ",
constant_chain, ".", to_member_name(type, mbr_idx), ";");
if (unroll_array)
{
// Member arrays in these blocks are always native arrays, unroll copy.
uint32_t len = to_array_size_literal(mbr_type);
for (uint32_t i = 0; i < len; i++)
{
statement(to_expression(stage_out_ptr_var_id), "[",
builtin_to_glsl(BuiltInInvocationId, StorageClassInput), "].",
to_member_name(ib_type, index), "[", i, "] = ",
constant_chain, ".", to_member_name(type, mbr_idx), "[", i, "];");
}
}
else
{
statement(to_expression(stage_out_ptr_var_id), "[",
builtin_to_glsl(BuiltInInvocationId, StorageClassInput), "].",
to_member_name(ib_type, index), " = ",
constant_chain, ".", to_member_name(type, mbr_idx), ";");
}
});
}
else