From d811258424e76d8b0047df3a8f3041a436d49ee7 Mon Sep 17 00:00:00 2001 From: Jamie Nicol Date: Fri, 13 Jun 2025 11:56:16 +0100 Subject: [PATCH] [naga msl-out] Add padding to end of structs if required --- naga/src/back/msl/writer.rs | 10 ++++++++++ naga/tests/out/msl/spv-quad-vert.msl | 1 + naga/tests/out/msl/spv-unnamed-gl-per-vertex.msl | 1 + naga/tests/out/msl/wgsl-6438-conflicting-idents.msl | 1 + naga/tests/out/msl/wgsl-access.msl | 2 ++ naga/tests/out/msl/wgsl-atomicCompareExchange.msl | 2 ++ naga/tests/out/msl/wgsl-atomicOps.msl | 2 ++ naga/tests/out/msl/wgsl-bounds-check-restrict.msl | 1 + naga/tests/out/msl/wgsl-bounds-check-zero.msl | 1 + naga/tests/out/msl/wgsl-constructors.msl | 1 + naga/tests/out/msl/wgsl-extra.msl | 1 + naga/tests/out/msl/wgsl-fragment-output.msl | 1 + naga/tests/out/msl/wgsl-int64.msl | 1 + naga/tests/out/msl/wgsl-interface.msl | 1 + naga/tests/out/msl/wgsl-interpolate.msl | 1 + naga/tests/out/msl/wgsl-interpolate_compat.msl | 1 + naga/tests/out/msl/wgsl-msl-vpt.msl | 2 ++ .../msl/wgsl-overrides-atomicCompareExchangeWeak.msl | 1 + naga/tests/out/msl/wgsl-padding.msl | 3 +++ tests/tests/wgpu-gpu/shader/struct_layout.rs | 5 ++--- 20 files changed, 36 insertions(+), 3 deletions(-) diff --git a/naga/src/back/msl/writer.rs b/naga/src/back/msl/writer.rs index 899a0ffb4..18a1de42f 100644 --- a/naga/src/back/msl/writer.rs +++ b/naga/src/back/msl/writer.rs @@ -4470,6 +4470,16 @@ impl Writer { } } } + if last_offset < span { + let pad = span - last_offset; + writeln!( + self.out, + "{}char _pad{}[{}];", + back::INDENT, + members.len(), + pad + )?; + } writeln!(self.out, "}};")?; } _ => { diff --git a/naga/tests/out/msl/spv-quad-vert.msl b/naga/tests/out/msl/spv-quad-vert.msl index 25bdd23df..349c1499b 100644 --- a/naga/tests/out/msl/spv-quad-vert.msl +++ b/naga/tests/out/msl/spv-quad-vert.msl @@ -12,6 +12,7 @@ struct gl_PerVertex { float gl_PointSize; type_3 gl_ClipDistance; type_3 gl_CullDistance; + char _pad4[4]; }; struct VertexOutput { metal::float2 member; diff --git a/naga/tests/out/msl/spv-unnamed-gl-per-vertex.msl b/naga/tests/out/msl/spv-unnamed-gl-per-vertex.msl index 83aea653d..947bfa6f9 100644 --- a/naga/tests/out/msl/spv-unnamed-gl-per-vertex.msl +++ b/naga/tests/out/msl/spv-unnamed-gl-per-vertex.msl @@ -12,6 +12,7 @@ struct type_4 { float member_1; type_3 member_2; type_3 member_3; + char _pad4[4]; }; void function( diff --git a/naga/tests/out/msl/wgsl-6438-conflicting-idents.msl b/naga/tests/out/msl/wgsl-6438-conflicting-idents.msl index 596ebad40..17287f628 100644 --- a/naga/tests/out/msl/wgsl-6438-conflicting-idents.msl +++ b/naga/tests/out/msl/wgsl-6438-conflicting-idents.msl @@ -7,6 +7,7 @@ using metal::uint; struct OurVertexShaderOutput { metal::float4 position; metal::float2 texcoord; + char _pad2[8]; }; struct vsInput { diff --git a/naga/tests/out/msl/wgsl-access.msl b/naga/tests/out/msl/wgsl-access.msl index a40a17ab6..f8ca0bf2f 100644 --- a/naga/tests/out/msl/wgsl-access.msl +++ b/naga/tests/out/msl/wgsl-access.msl @@ -16,6 +16,7 @@ struct GlobalConst { }; struct AlignedWrapper { int value; + char _pad1[4]; }; struct type_6 { metal::float2x2 inner[2]; @@ -35,6 +36,7 @@ struct Bar { char _pad4[4]; type_10 arr; type_11 data; + char _pad6[8]; }; struct Baz { metal::float3x2 m; diff --git a/naga/tests/out/msl/wgsl-atomicCompareExchange.msl b/naga/tests/out/msl/wgsl-atomicCompareExchange.msl index 2d5b6a9a2..72c418b9e 100644 --- a/naga/tests/out/msl/wgsl-atomicCompareExchange.msl +++ b/naga/tests/out/msl/wgsl-atomicCompareExchange.msl @@ -13,10 +13,12 @@ struct type_4 { struct _atomic_compare_exchange_resultSint4_ { int old_value; bool exchanged; + char _pad2[3]; }; struct _atomic_compare_exchange_resultUint4_ { uint old_value; bool exchanged; + char _pad2[3]; }; template diff --git a/naga/tests/out/msl/wgsl-atomicOps.msl b/naga/tests/out/msl/wgsl-atomicOps.msl index 860a5a3e5..63cab2a80 100644 --- a/naga/tests/out/msl/wgsl-atomicOps.msl +++ b/naga/tests/out/msl/wgsl-atomicOps.msl @@ -14,10 +14,12 @@ struct Struct { struct _atomic_compare_exchange_resultUint4_ { uint old_value; bool exchanged; + char _pad2[3]; }; struct _atomic_compare_exchange_resultSint4_ { int old_value; bool exchanged; + char _pad2[3]; }; template diff --git a/naga/tests/out/msl/wgsl-bounds-check-restrict.msl b/naga/tests/out/msl/wgsl-bounds-check-restrict.msl index c56c48f35..724e85655 100644 --- a/naga/tests/out/msl/wgsl-bounds-check-restrict.msl +++ b/naga/tests/out/msl/wgsl-bounds-check-restrict.msl @@ -18,6 +18,7 @@ struct Globals { metal::float4 v; metal::float3x4 m; type_4 d; + char _pad4[12]; }; float index_array( diff --git a/naga/tests/out/msl/wgsl-bounds-check-zero.msl b/naga/tests/out/msl/wgsl-bounds-check-zero.msl index 872a96808..1c95a20ac 100644 --- a/naga/tests/out/msl/wgsl-bounds-check-zero.msl +++ b/naga/tests/out/msl/wgsl-bounds-check-zero.msl @@ -24,6 +24,7 @@ struct Globals { metal::float4 v; metal::float3x4 m; type_4 d; + char _pad4[12]; }; float index_array( diff --git a/naga/tests/out/msl/wgsl-constructors.msl b/naga/tests/out/msl/wgsl-constructors.msl index a1d1c8a2e..61fe95442 100644 --- a/naga/tests/out/msl/wgsl-constructors.msl +++ b/naga/tests/out/msl/wgsl-constructors.msl @@ -7,6 +7,7 @@ using metal::uint; struct Foo { metal::float4 a; int b; + char _pad2[12]; }; struct type_6 { metal::float2x2 inner[1]; diff --git a/naga/tests/out/msl/wgsl-extra.msl b/naga/tests/out/msl/wgsl-extra.msl index 4d6bb568f..be46447e2 100644 --- a/naga/tests/out/msl/wgsl-extra.msl +++ b/naga/tests/out/msl/wgsl-extra.msl @@ -12,6 +12,7 @@ struct PushConstants { struct FragmentIn { metal::float4 color; uint primitive_index; + char _pad2[12]; }; struct main_Input { diff --git a/naga/tests/out/msl/wgsl-fragment-output.msl b/naga/tests/out/msl/wgsl-fragment-output.msl index c886fc885..0fe972f57 100644 --- a/naga/tests/out/msl/wgsl-fragment-output.msl +++ b/naga/tests/out/msl/wgsl-fragment-output.msl @@ -19,6 +19,7 @@ struct FragmentOutputVec2Scalar { float scalarf; int scalari; uint scalaru; + char _pad6[4]; }; struct main_vec4vec3_Output { diff --git a/naga/tests/out/msl/wgsl-int64.msl b/naga/tests/out/msl/wgsl-int64.msl index f1b4e1f64..2df5dfc95 100644 --- a/naga/tests/out/msl/wgsl-int64.msl +++ b/naga/tests/out/msl/wgsl-int64.msl @@ -21,6 +21,7 @@ struct UniformCompatible { metal::long3 val_i64_3_; metal::long4 val_i64_4_; ulong final_value; + char _pad12[24]; }; struct type_11 { ulong inner[2]; diff --git a/naga/tests/out/msl/wgsl-interface.msl b/naga/tests/out/msl/wgsl-interface.msl index 047873da9..fb1639955 100644 --- a/naga/tests/out/msl/wgsl-interface.msl +++ b/naga/tests/out/msl/wgsl-interface.msl @@ -7,6 +7,7 @@ using metal::uint; struct VertexOutput { metal::float4 position; float _varying; + char _pad2[12]; }; struct FragmentOutput { float depth; diff --git a/naga/tests/out/msl/wgsl-interpolate.msl b/naga/tests/out/msl/wgsl-interpolate.msl index c19005753..2dad749b5 100644 --- a/naga/tests/out/msl/wgsl-interpolate.msl +++ b/naga/tests/out/msl/wgsl-interpolate.msl @@ -18,6 +18,7 @@ struct FragmentInput { float perspective_centroid; float perspective_sample; float perspective_center; + char _pad12[4]; }; struct vert_mainOutput { diff --git a/naga/tests/out/msl/wgsl-interpolate_compat.msl b/naga/tests/out/msl/wgsl-interpolate_compat.msl index e386c07db..339e7c518 100644 --- a/naga/tests/out/msl/wgsl-interpolate_compat.msl +++ b/naga/tests/out/msl/wgsl-interpolate_compat.msl @@ -18,6 +18,7 @@ struct FragmentInput { float perspective_centroid; float perspective_sample; float perspective_center; + char _pad11[4]; }; struct vert_mainOutput { diff --git a/naga/tests/out/msl/wgsl-msl-vpt.msl b/naga/tests/out/msl/wgsl-msl-vpt.msl index df2afb775..b982f428d 100644 --- a/naga/tests/out/msl/wgsl-msl-vpt.msl +++ b/naga/tests/out/msl/wgsl-msl-vpt.msl @@ -13,11 +13,13 @@ struct VertexOutput { metal::float4 position; metal::float4 color; metal::float2 texcoord; + char _pad3[8]; }; struct VertexInput { metal::float4 position; metal::float3 normal; metal::float2 texcoord; + char _pad3[8]; }; float unpackFloat32_(uint b0, uint b1, uint b2, uint b3) { return as_type(b3 << 24 | b2 << 16 | b1 << 8 | b0); diff --git a/naga/tests/out/msl/wgsl-overrides-atomicCompareExchangeWeak.msl b/naga/tests/out/msl/wgsl-overrides-atomicCompareExchangeWeak.msl index d87190c59..48f09ee6d 100644 --- a/naga/tests/out/msl/wgsl-overrides-atomicCompareExchangeWeak.msl +++ b/naga/tests/out/msl/wgsl-overrides-atomicCompareExchangeWeak.msl @@ -7,6 +7,7 @@ using metal::uint; struct _atomic_compare_exchange_resultUint4_ { uint old_value; bool exchanged; + char _pad2[3]; }; template diff --git a/naga/tests/out/msl/wgsl-padding.msl b/naga/tests/out/msl/wgsl-padding.msl index ae11b7d16..23dbf8bc2 100644 --- a/naga/tests/out/msl/wgsl-padding.msl +++ b/naga/tests/out/msl/wgsl-padding.msl @@ -10,6 +10,7 @@ struct S { struct Test { S a; float b; + char _pad2[12]; }; struct type_2 { metal::float3 inner[2]; @@ -17,10 +18,12 @@ struct type_2 { struct Test2_ { type_2 a; float b; + char _pad2[12]; }; struct Test3_ { metal::float4x3 a; float b; + char _pad2[12]; }; struct vertex_Output { diff --git a/tests/tests/wgpu-gpu/shader/struct_layout.rs b/tests/tests/wgpu-gpu/shader/struct_layout.rs index cf57e2a04..e9378f401 100644 --- a/tests/tests/wgpu-gpu/shader/struct_layout.rs +++ b/tests/tests/wgpu-gpu/shader/struct_layout.rs @@ -211,7 +211,7 @@ fn create_struct_layout_tests(storage_type: InputStorageType) -> Vec // Test for https://github.com/gfx-rs/wgpu/issues/5262. // - // The struct is supposed to have a size of 32 and alignment of 16, but on metal, it has size 24. + // The struct is supposed to have a size of 32 and alignment of 16. for ty in ["f32", "u32", "i32"] { let header = format!("struct Inner {{ vec: vec3<{ty}>, scalar1: u32, scalar2: u32 }}"); let members = String::from("arr: array"); @@ -238,8 +238,7 @@ fn create_struct_layout_tests(storage_type: InputStorageType) -> Vec &input_values, &[0, 1, 2, 3, 4, 8, 9, 10, 11, 12], ) - .header(header) - .failures(Backends::METAL), + .header(header), ); }