fix(msl-out): emit and init struct member padding always

Resolves [`gfx-rs/wgpu`#4701](#4701).
gfx-rs · Nov 21, 2023 · fc12129 · fc12129
1 parent 7246226
commit fc12129
Show file tree

Hide file tree

Showing 3 changed files with 11 additions and 7 deletions.
diff --git a/naga/CHANGELOG.md b/naga/CHANGELOG.md
@@ -79,6 +79,7 @@ For changelogs after v0.14, see [the wgpu changelog](../CHANGELOG.md).
 
 - Add and fix minimum Metal version checks for optional functionality. ([#2486](https://github.com/gfx-rs/naga/pull/2486)) **@teoxoy**
 - Make varyings' struct members unique. ([#2521](https://github.com/gfx-rs/naga/pull/2521)) **@evahop**
+- Emit and init `struct` member padding always ([#4701](https://github.com/gfx-rs/wgpu/pull/4701)) **@ErichDonGubler**
 
 #### GLSL-OUT
 

diff --git a/naga/src/back/msl/writer.rs b/naga/src/back/msl/writer.rs
@@ -357,11 +357,6 @@ fn should_pack_struct_member(
     module: &crate::Module,
 ) -> Option<crate::Scalar> {
     let member = &members[index];
-    //Note: this is imperfect - the same structure can be used for host-shared
-    // things, where packed float would matter.
-    if member.binding.is_some() {
-        return None;
-    }
 
     let ty_inner = &module.types[member.ty].inner;
     let last_offset = member.offset + ty_inner.size(module.to_ctx());
@@ -3308,7 +3303,7 @@ impl<W: Write> Writer<W> {
                     let mut last_offset = 0;
                     for (index, member) in members.iter().enumerate() {
                         // quick and dirty way to figure out if we need this...
-                        if member.binding.is_none() && member.offset > last_offset {
+                        if member.offset > last_offset {
                             self.struct_member_pads.insert((handle, index as u32));
                             let pad = member.offset - last_offset;
                             writeln!(self.out, "{}char _pad{}[{}];", back::INDENT, index, pad)?;
@@ -4275,6 +4270,13 @@ impl<W: Write> Writer<W> {
                             if member_index != 0 {
                                 write!(self.out, ", ")?;
                             }
+                            // insert padding initialization, if needed
+                            if self
+                                .struct_member_pads
+                                .contains(&(arg.ty, member_index as u32))
+                            {
+                                write!(self.out, "{{}}, ")?;
+                            }
                             if let Some(crate::Binding::Location { .. }) = member.binding {
                                 write!(self.out, "{varyings_member_name}.")?;
                             }

diff --git a/naga/tests/out/msl/quad.msl b/naga/tests/out/msl/quad.msl
@@ -6,6 +6,7 @@ using metal::uint;
 
 struct VertexOutput {
     metal::float2 uv;
+    char _pad1[8];
     metal::float4 position;
 };
 constant float c_scale = 1.2;
@@ -23,7 +24,7 @@ vertex vert_mainOutput vert_main(
 ) {
     const auto pos = varyings.pos;
     const auto uv = varyings.uv;
-    const auto _tmp = VertexOutput {uv, metal::float4(c_scale * pos, 0.0, 1.0)};
+    const auto _tmp = VertexOutput {uv, {}, metal::float4(c_scale * pos, 0.0, 1.0)};
     return vert_mainOutput { _tmp.uv, _tmp.position };
 }