From 0b38e963791ddc4061d197ec1cae9e067c49b3d9 Mon Sep 17 00:00:00 2001
From: Teodor Tanasoaia <28601907+teoxoy@users.noreply.github.com>
Date: Sat, 20 Sep 2025 00:11:31 +0200
Subject: [PATCH] [msl] vertex pulling: fix unpacking `snorm16x2` & `snorm16x4`
 (#8249)

---
 cts_runner/test.lst                           |  1 +
 naga/src/back/msl/writer.rs                   | 30 +++++++++----------
 .../tests/out/msl/wgsl-msl-vpt-formats-x1.msl |  8 ++---
 .../tests/out/msl/wgsl-msl-vpt-formats-x2.msl |  8 ++---
 .../tests/out/msl/wgsl-msl-vpt-formats-x3.msl |  8 ++---
 .../tests/out/msl/wgsl-msl-vpt-formats-x4.msl |  8 ++---
 6 files changed, 32 insertions(+), 31 deletions(-)

diff --git a/cts_runner/test.lst b/cts_runner/test.lst
index 99f06512a..20fb875f3 100644
--- a/cts_runner/test.lst
+++ b/cts_runner/test.lst
@@ -125,6 +125,7 @@ webgpu:api,operation,rendering,color_target_state:blend_constant,setting:*
 webgpu:api,operation,rendering,depth:*
 webgpu:api,operation,rendering,draw:*
 webgpu:api,operation,shader_module,compilation_info:*
+webgpu:api,operation,vertex_state,correctness:non_zero_array_stride_and_attribute_offset:*
 // Likely due to https://github.com/gfx-rs/wgpu/issues/7357.
 fails-if(metal) webgpu:api,operation,uncapturederror:iff_uncaptured:*
 //FAIL: webgpu:shader,execution,expression,call,builtin,select:*
diff --git a/naga/src/back/msl/writer.rs b/naga/src/back/msl/writer.rs
index 6fa5c135e..0bd8d0ab3 100644
--- a/naga/src/back/msl/writer.rs
+++ b/naga/src/back/msl/writer.rs
@@ -5144,14 +5144,14 @@ template <typename A>
                 let name = self.namer.call("unpackSnorm16x2");
                 writeln!(
                     self.out,
-                    "metal::float2 {name}(metal::ushort b0, \
-                                          metal::ushort b1, \
-                                          metal::ushort b2, \
-                                          metal::ushort b3) {{"
+                    "metal::float2 {name}(uint b0, \
+                                          uint b1, \
+                                          uint b2, \
+                                          uint b3) {{"
                 )?;
                 writeln!(
                     self.out,
-                    "{}return metal::unpack_snorm2x16_to_float(b1 << 24 | b0 << 16 | b3 << 8 | b2);",
+                    "{}return metal::unpack_snorm2x16_to_float(b3 << 24 | b2 << 16 | b1 << 8 | b0);",
                     back::INDENT
                 )?;
                 writeln!(self.out, "}}")?;
@@ -5161,19 +5161,19 @@ template <typename A>
                 let name = self.namer.call("unpackSnorm16x4");
                 writeln!(
                     self.out,
-                    "metal::float4 {name}(metal::ushort b0, \
-                                          metal::ushort b1, \
-                                          metal::ushort b2, \
-                                          metal::ushort b3, \
-                                          metal::ushort b4, \
-                                          metal::ushort b5, \
-                                          metal::ushort b6, \
-                                          metal::ushort b7) {{"
+                    "metal::float4 {name}(uint b0, \
+                                          uint b1, \
+                                          uint b2, \
+                                          uint b3, \
+                                          uint b4, \
+                                          uint b5, \
+                                          uint b6, \
+                                          uint b7) {{"
                 )?;
                 writeln!(
                     self.out,
-                    "{}return metal::float4(metal::unpack_snorm2x16_to_float(b1 << 24 | b0 << 16 | b3 << 8 | b2), \
-                                            metal::unpack_snorm2x16_to_float(b5 << 24 | b4 << 16 | b7 << 8 | b6));",
+                    "{}return metal::float4(metal::unpack_snorm2x16_to_float(b3 << 24 | b2 << 16 | b1 << 8 | b0), \
+                                            metal::unpack_snorm2x16_to_float(b7 << 24 | b6 << 16 | b5 << 8 | b4));",
                     back::INDENT
                 )?;
                 writeln!(self.out, "}}")?;
diff --git a/naga/tests/out/msl/wgsl-msl-vpt-formats-x1.msl b/naga/tests/out/msl/wgsl-msl-vpt-formats-x1.msl
index b2a461c0b..766ffff6e 100644
--- a/naga/tests/out/msl/wgsl-msl-vpt-formats-x1.msl
+++ b/naga/tests/out/msl/wgsl-msl-vpt-formats-x1.msl
@@ -120,11 +120,11 @@ metal::float4 unpackUnorm16x4_(metal::ushort b0, metal::ushort b1, metal::ushort
 float unpackSnorm16_(metal::ushort b0, metal::ushort b1) {
     return metal::unpack_snorm2x16_to_float(b1 << 8 | b0).x;
 }
-metal::float2 unpackSnorm16x2_(metal::ushort b0, metal::ushort b1, metal::ushort b2, metal::ushort b3) {
-    return metal::unpack_snorm2x16_to_float(b1 << 24 | b0 << 16 | b3 << 8 | b2);
+metal::float2 unpackSnorm16x2_(uint b0, uint b1, uint b2, uint b3) {
+    return metal::unpack_snorm2x16_to_float(b3 << 24 | b2 << 16 | b1 << 8 | b0);
 }
-metal::float4 unpackSnorm16x4_(metal::ushort b0, metal::ushort b1, metal::ushort b2, metal::ushort b3, metal::ushort b4, metal::ushort b5, metal::ushort b6, metal::ushort b7) {
-    return metal::float4(metal::unpack_snorm2x16_to_float(b1 << 24 | b0 << 16 | b3 << 8 | b2), metal::unpack_snorm2x16_to_float(b5 << 24 | b4 << 16 | b7 << 8 | b6));
+metal::float4 unpackSnorm16x4_(uint b0, uint b1, uint b2, uint b3, uint b4, uint b5, uint b6, uint b7) {
+    return metal::float4(metal::unpack_snorm2x16_to_float(b3 << 24 | b2 << 16 | b1 << 8 | b0), metal::unpack_snorm2x16_to_float(b7 << 24 | b6 << 16 | b5 << 8 | b4));
 }
 float unpackFloat16_(metal::ushort b0, metal::ushort b1) {
     return float(as_type<half>(metal::ushort(b1 << 8 | b0)));
diff --git a/naga/tests/out/msl/wgsl-msl-vpt-formats-x2.msl b/naga/tests/out/msl/wgsl-msl-vpt-formats-x2.msl
index 46d1985d3..0559be74a 100644
--- a/naga/tests/out/msl/wgsl-msl-vpt-formats-x2.msl
+++ b/naga/tests/out/msl/wgsl-msl-vpt-formats-x2.msl
@@ -120,11 +120,11 @@ metal::float4 unpackUnorm16x4_(metal::ushort b0, metal::ushort b1, metal::ushort
 float unpackSnorm16_(metal::ushort b0, metal::ushort b1) {
     return metal::unpack_snorm2x16_to_float(b1 << 8 | b0).x;
 }
-metal::float2 unpackSnorm16x2_(metal::ushort b0, metal::ushort b1, metal::ushort b2, metal::ushort b3) {
-    return metal::unpack_snorm2x16_to_float(b1 << 24 | b0 << 16 | b3 << 8 | b2);
+metal::float2 unpackSnorm16x2_(uint b0, uint b1, uint b2, uint b3) {
+    return metal::unpack_snorm2x16_to_float(b3 << 24 | b2 << 16 | b1 << 8 | b0);
 }
-metal::float4 unpackSnorm16x4_(metal::ushort b0, metal::ushort b1, metal::ushort b2, metal::ushort b3, metal::ushort b4, metal::ushort b5, metal::ushort b6, metal::ushort b7) {
-    return metal::float4(metal::unpack_snorm2x16_to_float(b1 << 24 | b0 << 16 | b3 << 8 | b2), metal::unpack_snorm2x16_to_float(b5 << 24 | b4 << 16 | b7 << 8 | b6));
+metal::float4 unpackSnorm16x4_(uint b0, uint b1, uint b2, uint b3, uint b4, uint b5, uint b6, uint b7) {
+    return metal::float4(metal::unpack_snorm2x16_to_float(b3 << 24 | b2 << 16 | b1 << 8 | b0), metal::unpack_snorm2x16_to_float(b7 << 24 | b6 << 16 | b5 << 8 | b4));
 }
 float unpackFloat16_(metal::ushort b0, metal::ushort b1) {
     return float(as_type<half>(metal::ushort(b1 << 8 | b0)));
diff --git a/naga/tests/out/msl/wgsl-msl-vpt-formats-x3.msl b/naga/tests/out/msl/wgsl-msl-vpt-formats-x3.msl
index bd7471b38..7db698063 100644
--- a/naga/tests/out/msl/wgsl-msl-vpt-formats-x3.msl
+++ b/naga/tests/out/msl/wgsl-msl-vpt-formats-x3.msl
@@ -120,11 +120,11 @@ metal::float4 unpackUnorm16x4_(metal::ushort b0, metal::ushort b1, metal::ushort
 float unpackSnorm16_(metal::ushort b0, metal::ushort b1) {
     return metal::unpack_snorm2x16_to_float(b1 << 8 | b0).x;
 }
-metal::float2 unpackSnorm16x2_(metal::ushort b0, metal::ushort b1, metal::ushort b2, metal::ushort b3) {
-    return metal::unpack_snorm2x16_to_float(b1 << 24 | b0 << 16 | b3 << 8 | b2);
+metal::float2 unpackSnorm16x2_(uint b0, uint b1, uint b2, uint b3) {
+    return metal::unpack_snorm2x16_to_float(b3 << 24 | b2 << 16 | b1 << 8 | b0);
 }
-metal::float4 unpackSnorm16x4_(metal::ushort b0, metal::ushort b1, metal::ushort b2, metal::ushort b3, metal::ushort b4, metal::ushort b5, metal::ushort b6, metal::ushort b7) {
-    return metal::float4(metal::unpack_snorm2x16_to_float(b1 << 24 | b0 << 16 | b3 << 8 | b2), metal::unpack_snorm2x16_to_float(b5 << 24 | b4 << 16 | b7 << 8 | b6));
+metal::float4 unpackSnorm16x4_(uint b0, uint b1, uint b2, uint b3, uint b4, uint b5, uint b6, uint b7) {
+    return metal::float4(metal::unpack_snorm2x16_to_float(b3 << 24 | b2 << 16 | b1 << 8 | b0), metal::unpack_snorm2x16_to_float(b7 << 24 | b6 << 16 | b5 << 8 | b4));
 }
 float unpackFloat16_(metal::ushort b0, metal::ushort b1) {
     return float(as_type<half>(metal::ushort(b1 << 8 | b0)));
diff --git a/naga/tests/out/msl/wgsl-msl-vpt-formats-x4.msl b/naga/tests/out/msl/wgsl-msl-vpt-formats-x4.msl
index dba6657d6..d35045ffc 100644
--- a/naga/tests/out/msl/wgsl-msl-vpt-formats-x4.msl
+++ b/naga/tests/out/msl/wgsl-msl-vpt-formats-x4.msl
@@ -120,11 +120,11 @@ metal::float4 unpackUnorm16x4_(metal::ushort b0, metal::ushort b1, metal::ushort
 float unpackSnorm16_(metal::ushort b0, metal::ushort b1) {
     return metal::unpack_snorm2x16_to_float(b1 << 8 | b0).x;
 }
-metal::float2 unpackSnorm16x2_(metal::ushort b0, metal::ushort b1, metal::ushort b2, metal::ushort b3) {
-    return metal::unpack_snorm2x16_to_float(b1 << 24 | b0 << 16 | b3 << 8 | b2);
+metal::float2 unpackSnorm16x2_(uint b0, uint b1, uint b2, uint b3) {
+    return metal::unpack_snorm2x16_to_float(b3 << 24 | b2 << 16 | b1 << 8 | b0);
 }
-metal::float4 unpackSnorm16x4_(metal::ushort b0, metal::ushort b1, metal::ushort b2, metal::ushort b3, metal::ushort b4, metal::ushort b5, metal::ushort b6, metal::ushort b7) {
-    return metal::float4(metal::unpack_snorm2x16_to_float(b1 << 24 | b0 << 16 | b3 << 8 | b2), metal::unpack_snorm2x16_to_float(b5 << 24 | b4 << 16 | b7 << 8 | b6));
+metal::float4 unpackSnorm16x4_(uint b0, uint b1, uint b2, uint b3, uint b4, uint b5, uint b6, uint b7) {
+    return metal::float4(metal::unpack_snorm2x16_to_float(b3 << 24 | b2 << 16 | b1 << 8 | b0), metal::unpack_snorm2x16_to_float(b7 << 24 | b6 << 16 | b5 << 8 | b4));
 }
 float unpackFloat16_(metal::ushort b0, metal::ushort b1) {
     return float(as_type<half>(metal::ushort(b1 << 8 | b0)));