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

Tests run clean.

Browse Source
This commit is contained in:
Hans-Kristian Arntzen 2019-07-22 10:23:39 +02:00
parent 6c1f97b4a9
commit be2fccd837
14 changed files with 135 additions and 91 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

5
reference/opt/shaders-msl/asm/comp/block-name-alias-global.asm.comp
View File

@ -18,11 +18,12 @@ struct A_2
{
int a;
int b;
char _m0_final_padding[8];
};
struct A_3
{
/* FIXME: A padded struct is needed here. If you see this message, file a bug! */ A_2 Data[1024];
A_2 Data[1024];
};
struct B
@ -32,7 +33,7 @@ struct B
struct B_1
{
/* FIXME: A padded struct is needed here. If you see this message, file a bug! */ A_2 Data[1024];
A_2 Data[1024];
};
kernel void main0(device A_1& C1 [[buffer(0)]], constant A_3& C2 [[buffer(1)]], device B& C3 [[buffer(2)]], constant B_1& C4 [[buffer(3)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])

20
reference/opt/shaders-msl/comp/struct-packing.comp
View File

@ -3,12 +3,11 @@
using namespace metal;
typedef packed_float2 packed_rm_float2x3[3];
struct S0
{
float2 a[1];
float b;
char _m0_final_padding[4];
};
struct S1
@ -21,6 +20,7 @@ struct S2
{
float3 a[1];
float b;
char _m0_final_padding[12];
};
struct S3
@ -45,6 +45,7 @@ struct Content
S3 m3;
float m4;
S4 m3s[8];
char _m0_final_padding[8];
};
struct SSBO1
@ -58,10 +59,8 @@ struct SSBO1
float3x2 m3;
float2x2 m4;
float2x2 m5[9];
packed_rm_float2x3 m6[4][2];
char _m10_pad[8];
float3x2 m7;
char _m11_pad[8];
float3x2 m6[4][2];
float2x3 m7;
float array[1];
};
@ -69,6 +68,7 @@ struct S0_1
{
float4 a[1];
float b;
char _m0_final_padding[12];
};
struct S1_1
@ -81,6 +81,7 @@ struct S2_1
{
float3 a[1];
float b;
char _m0_final_padding[12];
};
struct S3_1
@ -92,6 +93,7 @@ struct S3_1
struct S4_1
{
float2 c;
char _m0_final_padding[8];
};
struct Content_1
@ -104,8 +106,8 @@ struct Content_1
S2_1 m2;
S3_1 m3;
float m4;
char _m8_pad[12];
/* FIXME: A padded struct is needed here. If you see this message, file a bug! */ S4_1 m3s[8];
char _m8_pad[8];
S4_1 m3s[8];
};
struct SSBO0
@ -142,6 +144,6 @@ kernel void main0(device SSBO1& ssbo_430 [[buffer(0)]], device SSBO0& ssbo_140 [
ssbo_430.content.m3s[5].c = _60.m3s[5].c;
ssbo_430.content.m3s[6].c = _60.m3s[6].c;
ssbo_430.content.m3s[7].c = _60.m3s[7].c;
ssbo_430.content.m1.a = ssbo_430.content.m3.a * float3x2(float2(ssbo_430.m6[1][1][0]), float2(ssbo_430.m6[1][1][1]), float2(ssbo_430.m6[1][1][2]));
ssbo_430.content.m1.a = transpose(ssbo_430.m6[1][1]) * ssbo_430.content.m3.a;
}

2
reference/opt/shaders-msl/flatten/rowmajor.flatten.vert
View File

@ -7,7 +7,7 @@ struct UBO
{
float4x4 uMVPR;
float4x4 uMVPC;
float2x4 uMVP;
float4x4 uMVP;
};
struct main0_out

8
reference/opt/shaders-msl/vert/packed_matrix.vert
View File

@ -3,12 +3,10 @@
using namespace metal;
typedef packed_float4 packed_rm_float4x3[3];
struct _15
{
packed_rm_float4x3 _m0;
packed_rm_float4x3 _m1;
float3x4 _m0;
float3x4 _m1;
};
struct _42
@ -41,7 +39,7 @@ vertex main0_out main0(main0_in in [[stage_in]], constant _15& _17 [[buffer(0)]]
{
main0_out out = {};
float4 _70 = _44._m0 * float4(float3(_44._m3) + (in.m_25.xyz * (_44._m6 + _44._m7)), 1.0);
out.m_72 = normalize(float4(in.m_25.xyz, 0.0) * float3x4(float4(_17._m1[0]), float4(_17._m1[1]), float4(_17._m1[2])));
out.m_72 = normalize(transpose(_17._m1) * float4(in.m_25.xyz, 0.0));
float4 _94 = _70;
_94.y = -_70.y;
out.gl_Position = _94;

12
reference/opt/shaders-msl/vert/read-from-row-major-array.vert
View File

@ -1,5 +1,3 @@
#pragma clang diagnostic ignored "-Wmissing-prototypes"
#include <metal_stdlib>
#include <simd/simd.h>
@ -7,7 +5,7 @@ using namespace metal;
struct Block
{
float2x3 var[3][4];
float3x4 var[3][4];
};
struct main0_out
@ -21,17 +19,11 @@ struct main0_in
float4 a_position [[attribute(0)]];
};
// Implementation of a conversion of matrix content from RowMajor to ColumnMajor organization.
float2x3 spvConvertFromRowMajor2x3(float2x3 m)
{
return float2x3(float3(m[0][0], m[0][2], m[1][1]), float3(m[0][1], m[1][0], m[1][2]));
}
vertex main0_out main0(main0_in in [[stage_in]], constant Block& _104 [[buffer(0)]])
{
main0_out out = {};
out.gl_Position = in.a_position;
out.v_vtxResult = ((float(abs(spvConvertFromRowMajor2x3(_104.var[0][0])[0].x - 2.0) < 0.0500000007450580596923828125) * float(abs(spvConvertFromRowMajor2x3(_104.var[0][0])[0].y - 6.0) < 0.0500000007450580596923828125)) * float(abs(spvConvertFromRowMajor2x3(_104.var[0][0])[0].z - (-6.0)) < 0.0500000007450580596923828125)) * ((float(abs(spvConvertFromRowMajor2x3(_104.var[0][0])[1].x) < 0.0500000007450580596923828125) * float(abs(spvConvertFromRowMajor2x3(_104.var[0][0])[1].y - 5.0) < 0.0500000007450580596923828125)) * float(abs(spvConvertFromRowMajor2x3(_104.var[0][0])[1].z - 5.0) < 0.0500000007450580596923828125));
out.v_vtxResult = ((float(abs(transpose(float2x3(_104.var[0][0][0].xy, _104.var[0][0][1].xy, _104.var[0][0][2].xy))[0].x - 2.0) < 0.0500000007450580596923828125) * float(abs(transpose(float2x3(_104.var[0][0][0].xy, _104.var[0][0][1].xy, _104.var[0][0][2].xy))[0].y - 6.0) < 0.0500000007450580596923828125)) * float(abs(transpose(float2x3(_104.var[0][0][0].xy, _104.var[0][0][1].xy, _104.var[0][0][2].xy))[0].z - (-6.0)) < 0.0500000007450580596923828125)) * ((float(abs(transpose(float2x3(_104.var[0][0][0].xy, _104.var[0][0][1].xy, _104.var[0][0][2].xy))[1].x) < 0.0500000007450580596923828125) * float(abs(transpose(float2x3(_104.var[0][0][0].xy, _104.var[0][0][1].xy, _104.var[0][0][2].xy))[1].y - 5.0) < 0.0500000007450580596923828125)) * float(abs(transpose(float2x3(_104.var[0][0][0].xy, _104.var[0][0][1].xy, _104.var[0][0][2].xy))[1].z - 5.0) < 0.0500000007450580596923828125));
return out;
}

63
reference/opt/shaders-msl/vulkan/comp/struct-packing-scalar.nocompat.invalid.vk.comp
View File

@ -4,37 +4,31 @@
using namespace metal;
typedef packed_float2 packed_float2x2[2];
typedef packed_float2 packed_rm_float2x3[3];
typedef packed_float3 packed_float2x3[2];
typedef packed_float3 packed_rm_float3x2[2];
struct S0
{
packed_float2 a[1];
float b;
packed_float b;
};
struct S1
{
packed_float3 a;
float b;
packed_float b;
};
struct S2
{
packed_float3 a[1];
float b;
packed_float b;
};
struct S3
{
packed_float2 a;
float b;
};
struct S4
{
float2 c;
packed_float b;
};
struct Content
@ -47,7 +41,6 @@ struct Content
S2 m2;
S3 m3;
float m4;
S4 m3s[8];
};
struct SSBO1
@ -61,8 +54,8 @@ struct SSBO1
float3x2 m3;
float2x2 m4;
float2x2 m5[9];
packed_rm_float2x3 m6[4][2];
float3x2 m7;
float3x2 m6[4][2];
packed_rm_float3x2 m7;
float array[1];
};
@ -70,6 +63,7 @@ struct S0_1
{
float4 a[1];
float b;
char _m0_final_padding[12];
};
struct S1_1
@ -82,6 +76,7 @@ struct S2_1
{
float3 a[1];
float b;
char _m0_final_padding[12];
};
struct S3_1
@ -90,11 +85,6 @@ struct S3_1
float b;
};
struct S4_1
{
float2 c;
};
struct Content_1
{
S0_1 m0s[1];
@ -105,8 +95,7 @@ struct Content_1
S2_1 m2;
S3_1 m3;
float m4;
char _m8_pad[12];
/* FIXME: A padded struct is needed here. If you see this message, file a bug! */ S4_1 m3s[8];
char _m0_final_padding[12];
};
struct SSBO0
@ -114,18 +103,14 @@ struct SSBO0
Content_1 content;
Content_1 content1[2];
Content_1 content2;
float2x2 m0;
char _m4_pad[16];
float2x2 m1;
char _m5_pad[16];
float2x4 m0;
float2x4 m1;
float2x3 m2[4];
float3x2 m3;
char _m7_pad[24];
float2x2 m4;
char _m8_pad[16];
float2x2 m5[9];
float2x3 m6[4][2];
float3x2 m7;
float3x4 m3;
float2x4 m4;
float2x4 m5[9];
float3x4 m6[4][2];
float2x3 m7;
float4 array[1];
};
@ -153,19 +138,11 @@ kernel void main0(device SSBO1& ssbo_scalar [[buffer(0)]], device SSBO0& ssbo_14
ssbo_scalar.content.m3.a = ssbo_140.content.m3.a;
ssbo_scalar.content.m3.b = ssbo_140.content.m3.b;
ssbo_scalar.content.m4 = ssbo_140.content.m4;
ssbo_scalar.content.m3s[0].c = ssbo_140.content.m3s[0].c;
ssbo_scalar.content.m3s[1].c = ssbo_140.content.m3s[1].c;
ssbo_scalar.content.m3s[2].c = ssbo_140.content.m3s[2].c;
ssbo_scalar.content.m3s[3].c = ssbo_140.content.m3s[3].c;
ssbo_scalar.content.m3s[4].c = ssbo_140.content.m3s[4].c;
ssbo_scalar.content.m3s[5].c = ssbo_140.content.m3s[5].c;
ssbo_scalar.content.m3s[6].c = ssbo_140.content.m3s[6].c;
ssbo_scalar.content.m3s[7].c = ssbo_140.content.m3s[7].c;
ssbo_scalar.content.m1.a = float2x3(float3(ssbo_scalar.m2[1][0]), float3(ssbo_scalar.m2[1][1])) * float2(ssbo_scalar.content.m0.a[0]);
ssbo_scalar.m0 = float2x2(float2(ssbo_scalar2.m1[0]), float2(ssbo_scalar2.m1[1]));
ssbo_scalar2.m1[0] = transpose(ssbo_scalar.m4)[0];
ssbo_scalar2.m1[1] = transpose(ssbo_scalar.m4)[1];
ssbo_scalar2.m2[0] = spvConvertFromRowMajor3x2(ssbo_scalar.m3)[0];
ssbo_scalar2.m2[1] = spvConvertFromRowMajor3x2(ssbo_scalar.m3)[1];
ssbo_scalar2.m1[0] = float2(ssbo_scalar.m4[0][0], ssbo_scalar.m4[1][0]));
ssbo_scalar2.m1[1] = float2(ssbo_scalar.m4[0][1], ssbo_scalar.m4[1][1]));
ssbo_scalar2.m2[0] = float3(ssbo_scalar.m3[0][0], ssbo_scalar.m3[1][0], ssbo_scalar.m3[2][0]));
ssbo_scalar2.m2[1] = float3(ssbo_scalar.m3[0][1], ssbo_scalar.m3[1][1], ssbo_scalar.m3[2][1]));
}

5
reference/shaders-msl/asm/comp/block-name-alias-global.asm.comp
View File

@ -18,11 +18,12 @@ struct A_2
{
int a;
int b;
char _m0_final_padding[8];
};
struct A_3
{
/* FIXME: A padded struct is needed here. If you see this message, file a bug! */ A_2 Data[1024];
A_2 Data[1024];
};
struct B
@ -32,7 +33,7 @@ struct B
struct B_1
{
/* FIXME: A padded struct is needed here. If you see this message, file a bug! */ A_2 Data[1024];
A_2 Data[1024];
};
kernel void main0(device A_1& C1 [[buffer(0)]], constant A_3& C2 [[buffer(1)]], device B& C3 [[buffer(2)]], constant B_1& C4 [[buffer(3)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])

20
reference/shaders-msl/comp/struct-packing.comp
View File

@ -3,12 +3,11 @@
using namespace metal;
typedef packed_float2 packed_rm_float2x3[3];
struct S0
{
float2 a[1];
float b;
char _m0_final_padding[4];
};
struct S1
@ -21,6 +20,7 @@ struct S2
{
float3 a[1];
float b;
char _m0_final_padding[12];
};
struct S3
@ -45,6 +45,7 @@ struct Content
S3 m3;
float m4;
S4 m3s[8];
char _m0_final_padding[8];
};
struct SSBO1
@ -58,10 +59,8 @@ struct SSBO1
float3x2 m3;
float2x2 m4;
float2x2 m5[9];
packed_rm_float2x3 m6[4][2];
char _m10_pad[8];
float3x2 m7;
char _m11_pad[8];
float3x2 m6[4][2];
float2x3 m7;
float array[1];
};
@ -69,6 +68,7 @@ struct S0_1
{
float4 a[1];
float b;
char _m0_final_padding[12];
};
struct S1_1
@ -81,6 +81,7 @@ struct S2_1
{
float3 a[1];
float b;
char _m0_final_padding[12];
};
struct S3_1
@ -92,6 +93,7 @@ struct S3_1
struct S4_1
{
float2 c;
char _m0_final_padding[8];
};
struct Content_1
@ -104,8 +106,8 @@ struct Content_1
S2_1 m2;
S3_1 m3;
float m4;
char _m8_pad[12];
/* FIXME: A padded struct is needed here. If you see this message, file a bug! */ S4_1 m3s[8];
char _m8_pad[8];
S4_1 m3s[8];
};
struct SSBO0
@ -142,6 +144,6 @@ kernel void main0(device SSBO1& ssbo_430 [[buffer(0)]], device SSBO0& ssbo_140 [
ssbo_430.content.m3s[5].c = _60.m3s[5].c;
ssbo_430.content.m3s[6].c = _60.m3s[6].c;
ssbo_430.content.m3s[7].c = _60.m3s[7].c;
ssbo_430.content.m1.a = ssbo_430.content.m3.a * float3x2(float2(ssbo_430.m6[1][1][0]), float2(ssbo_430.m6[1][1][1]), float2(ssbo_430.m6[1][1][2]));
ssbo_430.content.m1.a = transpose(ssbo_430.m6[1][1]) * ssbo_430.content.m3.a;
}

3
reference/shaders-msl/vulkan/comp/struct-packing-scalar.nocompat.invalid.vk.comp
View File

@ -63,6 +63,7 @@ struct S0_1
{
float4 a[1];
float b;
char _m0_final_padding[12];
};
struct S1_1
@ -75,6 +76,7 @@ struct S2_1
{
float3 a[1];
float b;
char _m0_final_padding[12];
};
struct S3_1
@ -93,6 +95,7 @@ struct Content_1
S2_1 m2;
S3_1 m3;
float m4;
char _m0_final_padding[12];
};
struct SSBO0

3
spirv_common.hpp
View File

@ -1420,7 +1420,8 @@ enum ExtendedDecorations
SPIRVCrossDecorationPhysicalTypePacked,
// The padding in bytes before declaring this struct member.
SPIRVCrossDecorationPadding,
// If used on a struct type, marks the target size of a struct.
SPIRVCrossDecorationPaddingTarget,
SPIRVCrossDecorationInterfaceMemberIndex,
SPIRVCrossDecorationInterfaceOrigID,

7
spirv_glsl.cpp
View File

@ -829,6 +829,9 @@ void CompilerGLSL::emit_struct(SPIRType &type)
emitted = true;
}
if (has_extended_decoration(type.self, SPIRVCrossDecorationPaddingTarget))
emit_struct_padding_target(type);
end_scope_decl();
if (emitted)
@ -9972,6 +9975,10 @@ void CompilerGLSL::emit_struct_member(const SPIRType &type, uint32_t member_type
variable_decl(membertype, to_member_name(type, index)), ";");
}
void CompilerGLSL::emit_struct_padding_target(const SPIRType &)
{
}
const char *CompilerGLSL::flags_to_qualifiers_glsl(const SPIRType &type, const Bitset &flags)
{
// GL_EXT_buffer_reference variables can be marked as restrict.

1
spirv_glsl.hpp
View File

@ -249,6 +249,7 @@ protected:
virtual std::string builtin_to_glsl(spv::BuiltIn builtin, spv::StorageClass storage);
virtual void emit_struct_member(const SPIRType &type, uint32_t member_type_id, uint32_t index,
const std::string &qualifier = "", uint32_t base_offset = 0);
virtual void emit_struct_padding_target(const SPIRType &type);
virtual std::string image_type_glsl(const SPIRType &type, uint32_t id = 0);
std::string constant_expression(const SPIRConstant &c);
std::string constant_op_expression(const SPIRConstantOp &cop);

74
spirv_msl.cpp
View File

@ -2420,6 +2420,27 @@ void CompilerMSL::mark_scalar_layout_structs(const SPIRType &type)
// and the next member will be placed at offset 12.
bool struct_is_misaligned = (spirv_offset % msl_alignment) != 0;
bool struct_is_too_large = spirv_offset + msl_size > spirv_offset_next;
uint32_t array_stride = 0;
bool struct_needs_explicit_padding = false;
// Verify that if a struct is used as an array that ArrayStride matches the effective size of the struct.
if (!mbr_type.array.empty())
{
array_stride = type_struct_member_array_stride(type, i);
uint32_t dimensions = type.array.size();
for (uint32_t dim = 1; dim < dimensions; dim++)
{
uint32_t array_size = to_array_size_literal(type, dim);
array_stride /= max(array_size, 1u);
}
// Set expected struct size based on ArrayStride.
struct_needs_explicit_padding = true;
// If struct size is larger than array stride, we might be able to fit, if we tightly pack.
if (get_declared_struct_size_msl(*struct_type) > array_stride)
struct_is_too_large = true;
}
if (struct_is_misaligned || struct_is_too_large)
{
@ -2433,6 +2454,25 @@ void CompilerMSL::mark_scalar_layout_structs(const SPIRType &type)
}
mark_scalar_layout_structs(*struct_type);
if (struct_needs_explicit_padding)
{
msl_size = get_declared_struct_size_msl(*struct_type, true, true);
if (array_stride < msl_size)
{
SPIRV_CROSS_THROW("Cannot express an array stride smaller than size of struct type.");
}
else
{
if (has_extended_decoration(struct_type->self, SPIRVCrossDecorationPaddingTarget))
{
if (array_stride != get_extended_decoration(struct_type->self, SPIRVCrossDecorationPaddingTarget))
SPIRV_CROSS_THROW("A struct is used with different array strides. Cannot express this in MSL.");
}
else
set_extended_decoration(struct_type->self, SPIRVCrossDecorationPaddingTarget, array_stride);
}
}
}
}
}
@ -2488,7 +2528,7 @@ void CompilerMSL::align_struct(SPIRType &ib_type, unordered_set<uint32_t> &align
// size rules, we'll pad to standard C-packing rules with a char[] array. If the member is farther
// away than C-packing, expects, add an inert padding member before the the member.
uint32_t padding_bytes = spirv_mbr_offset - aligned_msl_offset;
set_extended_member_decoration(ib_type_id, mbr_idx, SPIRVCrossDecorationPadding, padding_bytes);
set_extended_member_decoration(ib_type_id, mbr_idx, SPIRVCrossDecorationPaddingTarget, padding_bytes);
// Re-align as a sanity check that aligning post-padding matches up.
msl_offset += padding_bytes;
@ -6165,15 +6205,25 @@ void CompilerMSL::emit_struct_member(const SPIRType &type, uint32_t member_type_
const string &qualifier, uint32_t)
{
// If this member requires padding to maintain its declared offset, emit a dummy padding member before it.
if (has_extended_member_decoration(type.self, index, SPIRVCrossDecorationPadding))
if (has_extended_member_decoration(type.self, index, SPIRVCrossDecorationPaddingTarget))
{
uint32_t pad_len = get_extended_member_decoration(type.self, index, SPIRVCrossDecorationPadding);
statement("char _m", index, "_pad", "[", to_string(pad_len), "];");
uint32_t pad_len = get_extended_member_decoration(type.self, index, SPIRVCrossDecorationPaddingTarget);
statement("char _m", index, "_pad", "[", pad_len, "];");
}
statement(to_struct_member(type, member_type_id, index, qualifier));
}
void CompilerMSL::emit_struct_padding_target(const SPIRType &type)
{
uint32_t struct_size = get_declared_struct_size_msl(type, true, true);
uint32_t target_size = get_extended_decoration(type.self, SPIRVCrossDecorationPaddingTarget);
if (target_size < struct_size)
SPIRV_CROSS_THROW("Cannot pad with negative bytes.");
else if (target_size > struct_size)
statement("char _m0_final_padding[", target_size - struct_size, "];");
}
// Return a MSL qualifier for the specified function attribute member
string CompilerMSL::member_attribute_qualifier(const SPIRType &type, uint32_t index)
{
@ -8983,8 +9033,12 @@ uint32_t CompilerMSL::get_declared_struct_member_matrix_stride_msl(const SPIRTyp
has_member_decoration(type.self, index, DecorationRowMajor));
}
uint32_t CompilerMSL::get_declared_struct_size_msl(const SPIRType &struct_type) const
uint32_t CompilerMSL::get_declared_struct_size_msl(const SPIRType &struct_type, bool ignore_alignment, bool ignore_padding) const
{
// If we have a target size, that is the declared size as well.
if (!ignore_padding && has_extended_decoration(struct_type.self, SPIRVCrossDecorationPaddingTarget))
return get_extended_decoration(struct_type.self, SPIRVCrossDecorationPaddingTarget);
if (struct_type.member_types.empty())
return 0;
@ -8992,10 +9046,14 @@ uint32_t CompilerMSL::get_declared_struct_size_msl(const SPIRType &struct_type)
// In MSL, a struct's alignment is equal to the maximum alignment of any of its members.
uint32_t alignment = 1;
for (uint32_t i = 0; i < mbr_cnt; i++)
if (!ignore_alignment)
{
uint32_t mbr_alignment = get_declared_struct_member_alignment_msl(struct_type, i);
alignment = max(alignment, mbr_alignment);
for (uint32_t i = 0; i < mbr_cnt; i++)
{
uint32_t mbr_alignment = get_declared_struct_member_alignment_msl(struct_type, i);
alignment = max(alignment, mbr_alignment);
}
}
// Last member will always be matched to the final Offset decoration, but size of struct in MSL now depends

3
spirv_msl.hpp
View File

@ -423,6 +423,7 @@ protected:
const std::string &qualifier = "");
void emit_struct_member(const SPIRType &type, uint32_t member_type_id, uint32_t index,
const std::string &qualifier = "", uint32_t base_offset = 0) override;
void emit_struct_padding_target(const SPIRType &type) override;
std::string type_to_glsl(const SPIRType &type, uint32_t id = 0) override;
std::string image_type_glsl(const SPIRType &type, uint32_t id = 0) override;
std::string sampler_type(const SPIRType &type);
@ -531,7 +532,7 @@ protected:
const SPIRType &get_physical_member_type(const SPIRType &struct_type, uint32_t index) const;
uint32_t get_declared_struct_size_msl(const SPIRType &struct_type) const;
uint32_t get_declared_struct_size_msl(const SPIRType &struct_type, bool ignore_alignment = false, bool ignore_padding = false) const;
std::string to_component_argument(uint32_t id);
void align_struct(SPIRType &ib_type, std::unordered_set<uint32_t> &aligned_structs);