diff --git a/src/back/msl/keywords.rs b/src/back/msl/keywords.rs
index 670e2f85ed..f34b618db8 100644
--- a/src/back/msl/keywords.rs
+++ b/src/back/msl/keywords.rs
@@ -216,5 +216,4 @@ pub const RESERVED: &[&str] = &[
     "clamped_lod_e",
     super::writer::FREXP_FUNCTION,
     super::writer::MODF_FUNCTION,
-    super::writer::ISIGN_FUNCTION,
 ];
diff --git a/src/back/msl/writer.rs b/src/back/msl/writer.rs
index de0532f400..67ab887285 100644
--- a/src/back/msl/writer.rs
+++ b/src/back/msl/writer.rs
@@ -34,7 +34,6 @@ const RAY_QUERY_FUN_MAP_INTERSECTION: &str = "_map_intersection_type";
 
 pub(crate) const MODF_FUNCTION: &str = "naga_modf";
 pub(crate) const FREXP_FUNCTION: &str = "naga_frexp";
-pub(crate) const ISIGN_FUNCTION: &str = "naga_isign";
 
 /// Write the Metal name for a Naga numeric type: scalar, vector, or matrix.
 ///
@@ -1185,6 +1184,31 @@ impl<W: Write> Writer<W> {
         Ok(())
     }
 
+    /// Emit code for the sign(i32) expression.
+    ///
+    fn put_isign(
+        &mut self,
+        arg: Handle<crate::Expression>,
+        context: &ExpressionContext,
+    ) -> BackendResult {
+        write!(self.out, "{NAMESPACE}::select({NAMESPACE}::select(")?;
+        match context.resolve_type(arg) {
+            &crate::TypeInner::Vector { size, .. } => {
+                let size = back::vector_size_str(size);
+                write!(self.out, "int{size}(-1), int{size}(1)")?;
+            }
+            _ => {
+                write!(self.out, "-1, 1")?;
+            }
+        }
+        write!(self.out, ", (")?;
+        self.put_expression(arg, context, true)?;
+        write!(self.out, " > 0)), 0, (")?;
+        self.put_expression(arg, context, true)?;
+        write!(self.out, " == 0))")?;
+        Ok(())
+    }
+
     fn put_const_expression(
         &mut self,
         expr_handle: Handle<crate::Expression>,
@@ -1716,7 +1740,9 @@ impl<W: Write> Writer<W> {
                     Mf::Refract => "refract",
                     // computational
                     Mf::Sign => match arg_type.scalar_kind() {
-                        Some(crate::ScalarKind::Sint) => ISIGN_FUNCTION,
+                        Some(crate::ScalarKind::Sint) => {
+                            return self.put_isign(arg, context);
+                        }
                         _ => "sign",
                     },
                     Mf::Fma => "fma",
@@ -1821,7 +1847,7 @@ impl<W: Write> Writer<W> {
                     write!(self.out, "((")?;
                     self.put_expression(arg, context, false)?;
                     write!(self.out, ") * 57.295779513082322865)")?;
-                } else if fun == Mf::Modf || fun == Mf::Frexp || fun_name == ISIGN_FUNCTION {
+                } else if fun == Mf::Modf || fun == Mf::Frexp {
                     write!(self.out, "{fun_name}")?;
                     self.put_call_parameters(iter::once(arg), context)?;
                 } else {
@@ -2428,6 +2454,16 @@ impl<W: Write> Writer<W> {
                     crate::MathFunction::FindMsb => {
                         self.need_bake_expressions.insert(arg);
                     }
+                    crate::MathFunction::Sign => {
+                        // WGSL's `sign` function works also on signed ints, but Metal's only
+                        // works on floating points, so we emit inline code for integer `sign`
+                        // calls. But that code uses each argument 2 times (see `put_isign`),
+                        // so to avoid duplicated evaluation, we must bake the argument.
+                        let inner = context.resolve_type(expr_handle);
+                        if inner.scalar_kind() == Some(crate::ScalarKind::Sint) {
+                            self.need_bake_expressions.insert(arg);
+                        }
+                    }
                     _ => {}
                 }
             }
@@ -3096,7 +3132,6 @@ impl<W: Write> Writer<W> {
 
         self.write_type_defs(module)?;
         self.write_global_constants(module)?;
-        self.write_polyfills()?;
         self.write_functions(module, info, options, pipeline_options)
     }
 
@@ -4109,16 +4144,6 @@ impl<W: Write> Writer<W> {
         Ok(info)
     }
 
-    fn write_polyfills(&mut self) -> Result<(), Error> {
-        writeln!(
-            self.out,
-            "\ntemplate <typename T> inline T {ISIGN_FUNCTION}(T arg) {{
-    return {NAMESPACE}::select({NAMESPACE}::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}}"
-        )?;
-        Ok(())
-    }
-
     fn write_barrier(&mut self, flags: crate::Barrier, level: back::Level) -> BackendResult {
         // Note: OR-ring bitflags requires `__HAVE_MEMFLAG_OPERATORS__`,
         // so we try to avoid it here.
diff --git a/src/valid/expression.rs b/src/valid/expression.rs
index ddef1aaea0..502ce7420a 100644
--- a/src/valid/expression.rs
+++ b/src/valid/expression.rs
@@ -992,7 +992,7 @@ impl super::Validator {
                         }
                     }
                     Mf::Sign => {
-                        if arg1_ty.is_some() | arg2_ty.is_some() | arg3_ty.is_some() {
+                        if arg1_ty.is_some() || arg2_ty.is_some() || arg3_ty.is_some() {
                             return Err(ExpressionError::WrongArgumentCount(fun));
                         }
                         match *arg_ty {
diff --git a/tests/out/msl/access.msl b/tests/out/msl/access.msl
index d1fe75ee5d..e5d875dd19 100644
--- a/tests/out/msl/access.msl
+++ b/tests/out/msl/access.msl
@@ -58,10 +58,6 @@ struct type_26 {
     metal::float4 inner[2];
 };
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 void test_matrix_within_struct_accesses(
     constant Baz& baz
 ) {
diff --git a/tests/out/msl/array-in-ctor.msl b/tests/out/msl/array-in-ctor.msl
index e322333d61..9428cb1e74 100644
--- a/tests/out/msl/array-in-ctor.msl
+++ b/tests/out/msl/array-in-ctor.msl
@@ -11,10 +11,6 @@ struct Ah {
     type_1 inner;
 };
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 kernel void cs_main(
   device Ah const& ah [[user(fake0)]]
 ) {
diff --git a/tests/out/msl/array-in-function-return-type.msl b/tests/out/msl/array-in-function-return-type.msl
index 7475f01ff0..c2c2379cce 100644
--- a/tests/out/msl/array-in-function-return-type.msl
+++ b/tests/out/msl/array-in-function-return-type.msl
@@ -8,10 +8,6 @@ struct type_1 {
     float inner[2];
 };
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 type_1 ret_array(
 ) {
     return type_1 {1.0, 2.0};
diff --git a/tests/out/msl/atomicOps.msl b/tests/out/msl/atomicOps.msl
index 7bff77dff7..b7264e883d 100644
--- a/tests/out/msl/atomicOps.msl
+++ b/tests/out/msl/atomicOps.msl
@@ -12,10 +12,6 @@ struct Struct {
     type_2 atomic_arr;
 };
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 struct cs_mainInput {
 };
 kernel void cs_main(
diff --git a/tests/out/msl/binding-arrays.msl b/tests/out/msl/binding-arrays.msl
index b826460e01..10a5397910 100644
--- a/tests/out/msl/binding-arrays.msl
+++ b/tests/out/msl/binding-arrays.msl
@@ -17,10 +17,6 @@ struct FragmentIn {
     uint index;
 };
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 struct main_Input {
     uint index [[user(loc0), flat]];
 };
diff --git a/tests/out/msl/bitcast.msl b/tests/out/msl/bitcast.msl
index 75d6228a6c..a0cf093b7f 100644
--- a/tests/out/msl/bitcast.msl
+++ b/tests/out/msl/bitcast.msl
@@ -5,10 +5,6 @@
 using metal::uint;
 
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 kernel void main_(
 ) {
     metal::int2 i2_ = {};
diff --git a/tests/out/msl/bits.msl b/tests/out/msl/bits.msl
index 785f0adf60..2a00b6b843 100644
--- a/tests/out/msl/bits.msl
+++ b/tests/out/msl/bits.msl
@@ -5,10 +5,6 @@
 using metal::uint;
 
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 kernel void main_(
 ) {
     int i = {};
diff --git a/tests/out/msl/boids.msl b/tests/out/msl/boids.msl
index dcb92f51a5..1a81aaf684 100644
--- a/tests/out/msl/boids.msl
+++ b/tests/out/msl/boids.msl
@@ -28,10 +28,6 @@ struct Particles {
 };
 constant uint NUM_PARTICLES = 1500u;
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 struct main_Input {
 };
 kernel void main_(
diff --git a/tests/out/msl/bounds-check-image-restrict.msl b/tests/out/msl/bounds-check-image-restrict.msl
index 4b631593d1..9f94ef0a6e 100644
--- a/tests/out/msl/bounds-check-image-restrict.msl
+++ b/tests/out/msl/bounds-check-image-restrict.msl
@@ -5,10 +5,6 @@
 using metal::uint;
 
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 metal::float4 test_textureLoad_1d(
     int coords,
     int level,
diff --git a/tests/out/msl/bounds-check-image-rzsw.msl b/tests/out/msl/bounds-check-image-rzsw.msl
index 8d26f222a2..a93014cb27 100644
--- a/tests/out/msl/bounds-check-image-rzsw.msl
+++ b/tests/out/msl/bounds-check-image-rzsw.msl
@@ -11,10 +11,6 @@ struct DefaultConstructible {
 };
 
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 metal::float4 test_textureLoad_1d(
     int coords,
     int level,
diff --git a/tests/out/msl/bounds-check-restrict.msl b/tests/out/msl/bounds-check-restrict.msl
index 7c3c5ca6ee..49a232e706 100644
--- a/tests/out/msl/bounds-check-restrict.msl
+++ b/tests/out/msl/bounds-check-restrict.msl
@@ -20,10 +20,6 @@ struct Globals {
     type_4 d;
 };
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 float index_array(
     int i,
     device Globals const& globals,
diff --git a/tests/out/msl/bounds-check-zero-atomic.msl b/tests/out/msl/bounds-check-zero-atomic.msl
index a191700235..daaa079233 100644
--- a/tests/out/msl/bounds-check-zero-atomic.msl
+++ b/tests/out/msl/bounds-check-zero-atomic.msl
@@ -24,10 +24,6 @@ struct Globals {
     type_2 c;
 };
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 uint fetch_add_atomic(
     device Globals& globals,
     constant _mslBufferSizes& _buffer_sizes
diff --git a/tests/out/msl/bounds-check-zero.msl b/tests/out/msl/bounds-check-zero.msl
index 9dc96a6237..816983d98b 100644
--- a/tests/out/msl/bounds-check-zero.msl
+++ b/tests/out/msl/bounds-check-zero.msl
@@ -26,10 +26,6 @@ struct Globals {
     type_4 d;
 };
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 float index_array(
     int i,
     device Globals const& globals,
diff --git a/tests/out/msl/break-if.msl b/tests/out/msl/break-if.msl
index 18b8a8ff98..3a6f2e9bff 100644
--- a/tests/out/msl/break-if.msl
+++ b/tests/out/msl/break-if.msl
@@ -5,10 +5,6 @@
 using metal::uint;
 
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 void breakIfEmpty(
 ) {
     bool loop_init = true;
diff --git a/tests/out/msl/collatz.msl b/tests/out/msl/collatz.msl
index 04b902177a..88f9521a27 100644
--- a/tests/out/msl/collatz.msl
+++ b/tests/out/msl/collatz.msl
@@ -13,10 +13,6 @@ struct PrimeIndices {
     type_1 data;
 };
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 uint collatz_iterations(
     uint n_base
 ) {
diff --git a/tests/out/msl/constructors.msl b/tests/out/msl/constructors.msl
index 68face5ebb..b3d3b9dd43 100644
--- a/tests/out/msl/constructors.msl
+++ b/tests/out/msl/constructors.msl
@@ -30,10 +30,6 @@ constant type_10 cz6_ = type_10 {};
 constant Foo cz7_ = Foo {};
 constant type_11 cp3_ = type_11 {0, 1, 2, 3};
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 kernel void main_(
 ) {
     Foo foo = {};
diff --git a/tests/out/msl/control-flow.msl b/tests/out/msl/control-flow.msl
index 75312986e7..be396e23a8 100644
--- a/tests/out/msl/control-flow.msl
+++ b/tests/out/msl/control-flow.msl
@@ -5,10 +5,6 @@
 using metal::uint;
 
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 void switch_default_break(
     int i
 ) {
diff --git a/tests/out/msl/do-while.msl b/tests/out/msl/do-while.msl
index 8f6c56388e..4f928d1f3c 100644
--- a/tests/out/msl/do-while.msl
+++ b/tests/out/msl/do-while.msl
@@ -5,10 +5,6 @@
 using metal::uint;
 
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 void fb1_(
     thread bool& cond
 ) {
diff --git a/tests/out/msl/dualsource.msl b/tests/out/msl/dualsource.msl
index c5a49ce558..92b1909f8b 100644
--- a/tests/out/msl/dualsource.msl
+++ b/tests/out/msl/dualsource.msl
@@ -9,10 +9,6 @@ struct FragmentOutput {
     metal::float4 mask;
 };
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 struct main_Input {
 };
 struct main_Output {
diff --git a/tests/out/msl/empty-global-name.msl b/tests/out/msl/empty-global-name.msl
index 4c64c79aae..f456649bc6 100644
--- a/tests/out/msl/empty-global-name.msl
+++ b/tests/out/msl/empty-global-name.msl
@@ -8,10 +8,6 @@ struct type_1 {
     int member;
 };
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 void function(
     device type_1& unnamed
 ) {
diff --git a/tests/out/msl/empty.msl b/tests/out/msl/empty.msl
index 724e63aa0a..4f8bf9f5e9 100644
--- a/tests/out/msl/empty.msl
+++ b/tests/out/msl/empty.msl
@@ -5,10 +5,6 @@
 using metal::uint;
 
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 kernel void main_(
 ) {
     return;
diff --git a/tests/out/msl/extra.msl b/tests/out/msl/extra.msl
index 40272282cd..8288dfad92 100644
--- a/tests/out/msl/extra.msl
+++ b/tests/out/msl/extra.msl
@@ -14,10 +14,6 @@ struct FragmentIn {
     uint primitive_index;
 };
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 struct main_Input {
     metal::float4 color [[user(loc0), center_perspective]];
 };
diff --git a/tests/out/msl/fragment-output.msl b/tests/out/msl/fragment-output.msl
index 6ed78d6bb1..4d25809e4f 100644
--- a/tests/out/msl/fragment-output.msl
+++ b/tests/out/msl/fragment-output.msl
@@ -21,10 +21,6 @@ struct FragmentOutputVec2Scalar {
     uint scalaru;
 };
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 struct main_vec4vec3_Output {
     metal::float4 vec4f [[color(0)]];
     metal::int4 vec4i [[color(1)]];
diff --git a/tests/out/msl/functions.msl b/tests/out/msl/functions.msl
index abd1ff5718..31f57c656e 100644
--- a/tests/out/msl/functions.msl
+++ b/tests/out/msl/functions.msl
@@ -5,10 +5,6 @@
 using metal::uint;
 
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 metal::float2 test_fma(
 ) {
     metal::float2 a = metal::float2(2.0, 2.0);
diff --git a/tests/out/msl/globals.msl b/tests/out/msl/globals.msl
index 24ccdb73fa..d9142c1990 100644
--- a/tests/out/msl/globals.msl
+++ b/tests/out/msl/globals.msl
@@ -33,10 +33,6 @@ struct type_15 {
 };
 constant bool Foo_1 = true;
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 void test_msl_packed_vec3_as_arg(
     metal::float3 arg
 ) {
diff --git a/tests/out/msl/image.msl b/tests/out/msl/image.msl
index d59862f9b8..e390c2e0fc 100644
--- a/tests/out/msl/image.msl
+++ b/tests/out/msl/image.msl
@@ -5,10 +5,6 @@
 using metal::uint;
 
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 struct main_Input {
 };
 kernel void main_(
diff --git a/tests/out/msl/interface.msl b/tests/out/msl/interface.msl
index 711e021743..c9b93d86a6 100644
--- a/tests/out/msl/interface.msl
+++ b/tests/out/msl/interface.msl
@@ -23,10 +23,6 @@ struct Input2_ {
     uint index;
 };
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 struct vertex_Input {
     uint color [[attribute(10)]];
 };
diff --git a/tests/out/msl/interpolate.msl b/tests/out/msl/interpolate.msl
index 999edbd9e2..5d8b67111e 100644
--- a/tests/out/msl/interpolate.msl
+++ b/tests/out/msl/interpolate.msl
@@ -15,10 +15,6 @@ struct FragmentInput {
     float perspective_sample;
 };
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 struct vert_mainOutput {
     metal::float4 position [[position]];
     uint _flat [[user(loc0), flat]];
diff --git a/tests/out/msl/math-functions.msl b/tests/out/msl/math-functions.msl
index 7476c51f93..dccb90ad6c 100644
--- a/tests/out/msl/math-functions.msl
+++ b/tests/out/msl/math-functions.msl
@@ -55,10 +55,6 @@ __frexp_result_vec4_f32_ naga_frexp(metal::float4 arg) {
     return __frexp_result_vec4_f32_{ fract, other };
 }
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 fragment void main_(
 ) {
     metal::float4 v = metal::float4(0.0);
@@ -68,8 +64,9 @@ fragment void main_(
     metal::float4 d = ((v) * 0.017453292519943295474);
     metal::float4 e = metal::saturate(v);
     metal::float4 g = metal::refract(v, v, 1.0);
-    int sign_a = naga_isign(-1);
-    metal::int4 sign_b = naga_isign(metal::int4(-1));
+    int sign_a = metal::select(metal::select(-1, 1, (-1 > 0)), 0, (-1 == 0));
+    metal::int4 _e12 = metal::int4(-1);
+    metal::int4 sign_b = metal::select(metal::select(int4(-1), int4(1), (_e12 > 0)), 0, (_e12 == 0));
     float sign_c = metal::sign(-1.0);
     metal::float4 sign_d = metal::sign(metal::float4(-1.0));
     int const_dot = ( + metal::int2 {}.x * metal::int2 {}.x + metal::int2 {}.y * metal::int2 {}.y);
diff --git a/tests/out/msl/operators.msl b/tests/out/msl/operators.msl
index 76ef5ebbe2..de3e990db2 100644
--- a/tests/out/msl/operators.msl
+++ b/tests/out/msl/operators.msl
@@ -9,10 +9,6 @@ constant metal::float4 v_f32_zero = metal::float4(0.0, 0.0, 0.0, 0.0);
 constant metal::float4 v_f32_half = metal::float4(0.5, 0.5, 0.5, 0.5);
 constant metal::int4 v_i32_one = metal::int4(1, 1, 1, 1);
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 metal::float4 builtins(
 ) {
     int s1_ = true ? 1 : 0;
diff --git a/tests/out/msl/padding.msl b/tests/out/msl/padding.msl
index 0aa7659414..4d99bb4c4c 100644
--- a/tests/out/msl/padding.msl
+++ b/tests/out/msl/padding.msl
@@ -23,10 +23,6 @@ struct Test3_ {
     float b;
 };
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 struct vertex_Output {
     metal::float4 member [[position]];
 };
diff --git a/tests/out/msl/policy-mix.msl b/tests/out/msl/policy-mix.msl
index 2b05111c59..f6a4fe5d6d 100644
--- a/tests/out/msl/policy-mix.msl
+++ b/tests/out/msl/policy-mix.msl
@@ -32,10 +32,6 @@ struct type_9 {
     metal::float4 inner[2];
 };
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 metal::float4 mock_function(
     metal::int2 c,
     int i,
diff --git a/tests/out/msl/quad-vert.msl b/tests/out/msl/quad-vert.msl
index 370bfc8036..d322cbcb9b 100644
--- a/tests/out/msl/quad-vert.msl
+++ b/tests/out/msl/quad-vert.msl
@@ -18,10 +18,6 @@ struct type_9 {
     metal::float4 gl_Position;
 };
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 void main_1(
     thread metal::float2& v_uv,
     thread metal::float2& a_uv_1,
diff --git a/tests/out/msl/quad.msl b/tests/out/msl/quad.msl
index 368c89d179..5fa5788aac 100644
--- a/tests/out/msl/quad.msl
+++ b/tests/out/msl/quad.msl
@@ -10,10 +10,6 @@ struct VertexOutput {
 };
 constant float c_scale = 1.2;
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 struct vert_mainInput {
     metal::float2 pos [[attribute(0)]];
     metal::float2 uv [[attribute(1)]];
diff --git a/tests/out/msl/ray-query.msl b/tests/out/msl/ray-query.msl
index f773cade66..0d4560f313 100644
--- a/tests/out/msl/ray-query.msl
+++ b/tests/out/msl/ray-query.msl
@@ -41,10 +41,6 @@ struct RayDesc {
     metal::float3 dir;
 };
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 metal::float3 get_torus_normal(
     metal::float3 world_point,
     RayIntersection intersection
diff --git a/tests/out/msl/resource-binding-map.msl b/tests/out/msl/resource-binding-map.msl
index f53f6c8423..b4a53d97b5 100644
--- a/tests/out/msl/resource-binding-map.msl
+++ b/tests/out/msl/resource-binding-map.msl
@@ -11,10 +11,6 @@ struct DefaultConstructible {
 };
 
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 struct entry_point_oneInput {
 };
 struct entry_point_oneOutput {
diff --git a/tests/out/msl/shadow.msl b/tests/out/msl/shadow.msl
index 0c06512c03..53f320344a 100644
--- a/tests/out/msl/shadow.msl
+++ b/tests/out/msl/shadow.msl
@@ -33,10 +33,6 @@ struct type_7 {
 constant metal::float3 c_ambient = metal::float3(0.05, 0.05, 0.05);
 constant uint c_max_lights = 10u;
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 float fetch_shadow(
     uint light_id,
     metal::float4 homogeneous_coords,
diff --git a/tests/out/msl/skybox.msl b/tests/out/msl/skybox.msl
index cacc687a9a..7b10ea23e7 100644
--- a/tests/out/msl/skybox.msl
+++ b/tests/out/msl/skybox.msl
@@ -13,10 +13,6 @@ struct Data {
     metal::float4x4 view;
 };
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 struct vs_mainInput {
 };
 struct vs_mainOutput {
diff --git a/tests/out/msl/standard.msl b/tests/out/msl/standard.msl
index f1d5e10d0b..f02243eaac 100644
--- a/tests/out/msl/standard.msl
+++ b/tests/out/msl/standard.msl
@@ -5,10 +5,6 @@
 using metal::uint;
 
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 bool test_any_and_all_for_bool(
 ) {
     return true;
diff --git a/tests/out/msl/texture-arg.msl b/tests/out/msl/texture-arg.msl
index b11d2f9caa..5fb9b25649 100644
--- a/tests/out/msl/texture-arg.msl
+++ b/tests/out/msl/texture-arg.msl
@@ -5,10 +5,6 @@
 using metal::uint;
 
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 metal::float4 test(
     metal::texture2d<float, metal::access::sample> Passed_Texture,
     metal::sampler Passed_Sampler
diff --git a/tests/out/msl/workgroup-uniform-load.msl b/tests/out/msl/workgroup-uniform-load.msl
index 28a0df7ac5..37a8781739 100644
--- a/tests/out/msl/workgroup-uniform-load.msl
+++ b/tests/out/msl/workgroup-uniform-load.msl
@@ -9,10 +9,6 @@ struct type_2 {
 };
 constant uint SIZE = 128u;
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 struct test_workgroupUniformLoadInput {
 };
 kernel void test_workgroupUniformLoad(
diff --git a/tests/out/msl/workgroup-var-init.msl b/tests/out/msl/workgroup-var-init.msl
index ba495894fd..ac300d4337 100644
--- a/tests/out/msl/workgroup-var-init.msl
+++ b/tests/out/msl/workgroup-var-init.msl
@@ -19,10 +19,6 @@ struct WStruct {
     type_4 atom_arr;
 };
 
-template <typename T> inline T naga_isign(T arg) {
-    return metal::select(metal::select(T(-1), T(1), (arg > 0)), 0, (arg == 0));
-}
-
 kernel void main_(
   metal::uint3 __local_invocation_id [[thread_position_in_threadgroup]]
 , threadgroup WStruct& w_mem