tint: Implement signed-int overloads of sign()

Bug: tint:1581
Fixed: tint:1782
Change-Id: Ia029bf9d1ce1d978c5cabc3016cb8ad1b4bac06a
Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/113243
Reviewed-by: Dan Sinclair <dsinclair@chromium.org>
Commit-Queue: Ben Clayton <bclayton@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>

src/tint/intrinsics.def

@@ -548,8 +548,8 @@ fn pow<N: num, T: f32_f16>(vec<N, T>, vec<N, T>) -> vec<N, T>
 @const("select_bool") fn select<T: scalar>(T, T, bool) -> T
 @const("select_bool") fn select<T: scalar, N: num>(vec<N, T>, vec<N, T>, bool) -> vec<N, T>
 @const("select_boolvec") fn select<N: num, T: scalar>(vec<N, T>, vec<N, T>, vec<N, bool>) -> vec<N, T>
-@const fn sign<T: fa_f32_f16>(T) -> T
+@const fn sign<T: fia_fi32_f16>(T) -> T
-@const fn sign<N: num, T: fa_f32_f16>(vec<N, T>) -> vec<N, T>
+@const fn sign<N: num, T: fia_fi32_f16>(vec<N, T>) -> vec<N, T>
 @const fn sin<T: fa_f32_f16>(T) -> T
 @const fn sin<N: num, T: fa_f32_f16>(vec<N, T>) -> vec<N, T>
 @const fn sinh<T: fa_f32_f16>(T) -> T

src/tint/resolver/const_eval.cc

@@ -3298,7 +3298,7 @@ ConstEval::Result ConstEval::sign(const sem::Type* ty,
             }
             return CreateElement(builder, source, c0->Type(), result);
         };
-        return Dispatch_fa_f32_f16(create, c0);
+        return Dispatch_fia_fi32_f16(create, c0);
     };
     return TransformElements(builder, ty, transform, args[0]);
 }

src/tint/resolver/const_eval_builtin_test.cc

@@ -2200,27 +2200,40 @@ INSTANTIATE_TEST_SUITE_P(  //
 
 template <typename T>
 std::vector<Case> SignCases() {
-    return {
+    std::vector<Case> cases = {
-        C({-T(1)}, -T(1)),
-        C({-T(0.5)}, -T(1)),
         C({T(0)}, T(0)),
         C({-T(0)}, T(0)),
-        C({T(0.5)}, T(1)),
+
+        C({-T(1)}, -T(1)),
+        C({-T(10)}, -T(1)),
+        C({-T(100)}, -T(1)),
         C({T(1)}, T(1)),
+        C({T(10)}, T(1)),
+        C({T(100)}, T(1)),
 
         C({T::Highest()}, T(1.0)),
         C({T::Lowest()}, -T(1.0)),
 
         // Vector tests
-        C({Vec(-T(0.5), T(0), T(0.5))}, Vec(-T(1.0), T(0.0), T(1.0))),
         C({Vec(T::Highest(), T::Lowest())}, Vec(T(1.0), -T(1.0))),
     };
+
+    ConcatIntoIf<IsFloatingPoint<T>>(
+        cases, std::vector<Case>{
+                   C({-T(0.5)}, -T(1)),
+                   C({T(0.5)}, T(1)),
+                   C({Vec(-T(0.5), T(0), T(0.5))}, Vec(-T(1.0), T(0.0), T(1.0))),
+               });
+
+    return cases;
 }
 INSTANTIATE_TEST_SUITE_P(  //
     Sign,
     ResolverConstEvalBuiltinTest,
     testing::Combine(testing::Values(sem::BuiltinType::kSign),
-                     testing::ValuesIn(Concat(SignCases<AFloat>(),  //
+                     testing::ValuesIn(Concat(SignCases<AInt>(),  //
+                                              SignCases<i32>(),
+                                              SignCases<AFloat>(),
                                               SignCases<f32>(),
                                               SignCases<f16>()))));
 

src/tint/resolver/intrinsic_table.inl

@@ -8194,12 +8194,12 @@ constexpr TemplateTypeInfo kTemplateTypes[] = {
   {
     /* [28] */
     /* name */ "T",
-    /* matcher index */ 64,
+    /* matcher index */ 60,
   },
   {
     /* [29] */
     /* name */ "T",
-    /* matcher index */ 60,
+    /* matcher index */ 64,
   },
   {
     /* [30] */
@@ -11369,7 +11369,7 @@ constexpr OverloadInfo kOverloads[] = {
     /* num parameters */ 2,
     /* num template types */ 1,
     /* num template numbers */ 0,
-    /* template types */ &kTemplateTypes[28],
+    /* template types */ &kTemplateTypes[29],
     /* template numbers */ &kTemplateNumbers[10],
     /* parameters */ &kParameters[734],
     /* return matcher indices */ &kMatcherIndices[3],
@@ -11381,7 +11381,7 @@ constexpr OverloadInfo kOverloads[] = {
     /* num parameters */ 2,
     /* num template types */ 1,
     /* num template numbers */ 1,
-    /* template types */ &kTemplateTypes[28],
+    /* template types */ &kTemplateTypes[29],
     /* template numbers */ &kTemplateNumbers[4],
     /* parameters */ &kParameters[736],
     /* return matcher indices */ &kMatcherIndices[30],
@@ -11417,7 +11417,7 @@ constexpr OverloadInfo kOverloads[] = {
     /* num parameters */ 2,
     /* num template types */ 1,
     /* num template numbers */ 0,
-    /* template types */ &kTemplateTypes[28],
+    /* template types */ &kTemplateTypes[29],
     /* template numbers */ &kTemplateNumbers[10],
     /* parameters */ &kParameters[742],
     /* return matcher indices */ &kMatcherIndices[3],
@@ -11429,7 +11429,7 @@ constexpr OverloadInfo kOverloads[] = {
     /* num parameters */ 2,
     /* num template types */ 1,
     /* num template numbers */ 1,
-    /* template types */ &kTemplateTypes[28],
+    /* template types */ &kTemplateTypes[29],
     /* template numbers */ &kTemplateNumbers[4],
     /* parameters */ &kParameters[744],
     /* return matcher indices */ &kMatcherIndices[30],
@@ -12941,7 +12941,7 @@ constexpr OverloadInfo kOverloads[] = {
     /* num parameters */ 1,
     /* num template types */ 1,
     /* num template numbers */ 0,
-    /* template types */ &kTemplateTypes[23],
+    /* template types */ &kTemplateTypes[28],
     /* template numbers */ &kTemplateNumbers[10],
     /* parameters */ &kParameters[892],
     /* return matcher indices */ &kMatcherIndices[3],
@@ -12953,7 +12953,7 @@ constexpr OverloadInfo kOverloads[] = {
     /* num parameters */ 1,
     /* num template types */ 1,
     /* num template numbers */ 1,
-    /* template types */ &kTemplateTypes[23],
+    /* template types */ &kTemplateTypes[28],
     /* template numbers */ &kTemplateNumbers[4],
     /* parameters */ &kParameters[893],
     /* return matcher indices */ &kMatcherIndices[30],
@@ -13229,7 +13229,7 @@ constexpr OverloadInfo kOverloads[] = {
     /* num parameters */ 1,
     /* num template types */ 1,
     /* num template numbers */ 0,
-    /* template types */ &kTemplateTypes[28],
+    /* template types */ &kTemplateTypes[29],
     /* template numbers */ &kTemplateNumbers[10],
     /* parameters */ &kParameters[950],
     /* return matcher indices */ &kMatcherIndices[3],
@@ -13241,7 +13241,7 @@ constexpr OverloadInfo kOverloads[] = {
     /* num parameters */ 1,
     /* num template types */ 1,
     /* num template numbers */ 1,
-    /* template types */ &kTemplateTypes[28],
+    /* template types */ &kTemplateTypes[29],
     /* template numbers */ &kTemplateNumbers[4],
     /* parameters */ &kParameters[951],
     /* return matcher indices */ &kMatcherIndices[30],
@@ -13253,7 +13253,7 @@ constexpr OverloadInfo kOverloads[] = {
     /* num parameters */ 1,
     /* num template types */ 1,
     /* num template numbers */ 0,
-    /* template types */ &kTemplateTypes[29],
+    /* template types */ &kTemplateTypes[28],
     /* template numbers */ &kTemplateNumbers[10],
     /* parameters */ &kParameters[952],
     /* return matcher indices */ &kMatcherIndices[3],
@@ -13265,7 +13265,7 @@ constexpr OverloadInfo kOverloads[] = {
     /* num parameters */ 1,
     /* num template types */ 1,
     /* num template numbers */ 1,
-    /* template types */ &kTemplateTypes[29],
+    /* template types */ &kTemplateTypes[28],
     /* template numbers */ &kTemplateNumbers[4],
     /* parameters */ &kParameters[953],
     /* return matcher indices */ &kMatcherIndices[30],
@@ -13277,7 +13277,7 @@ constexpr OverloadInfo kOverloads[] = {
     /* num parameters */ 2,
     /* num template types */ 1,
     /* num template numbers */ 0,
-    /* template types */ &kTemplateTypes[28],
+    /* template types */ &kTemplateTypes[29],
     /* template numbers */ &kTemplateNumbers[10],
     /* parameters */ &kParameters[726],
     /* return matcher indices */ &kMatcherIndices[3],
@@ -13289,7 +13289,7 @@ constexpr OverloadInfo kOverloads[] = {
     /* num parameters */ 2,
     /* num template types */ 1,
     /* num template numbers */ 1,
-    /* template types */ &kTemplateTypes[28],
+    /* template types */ &kTemplateTypes[29],
     /* template numbers */ &kTemplateNumbers[4],
     /* parameters */ &kParameters[728],
     /* return matcher indices */ &kMatcherIndices[30],
@@ -13445,7 +13445,7 @@ constexpr OverloadInfo kOverloads[] = {
     /* num parameters */ 2,
     /* num template types */ 1,
     /* num template numbers */ 0,
-    /* template types */ &kTemplateTypes[28],
+    /* template types */ &kTemplateTypes[29],
     /* template numbers */ &kTemplateNumbers[10],
     /* parameters */ &kParameters[774],
     /* return matcher indices */ &kMatcherIndices[3],
@@ -13457,7 +13457,7 @@ constexpr OverloadInfo kOverloads[] = {
     /* num parameters */ 2,
     /* num template types */ 1,
     /* num template numbers */ 1,
-    /* template types */ &kTemplateTypes[28],
+    /* template types */ &kTemplateTypes[29],
     /* template numbers */ &kTemplateNumbers[4],
     /* parameters */ &kParameters[776],
     /* return matcher indices */ &kMatcherIndices[30],
@@ -13469,7 +13469,7 @@ constexpr OverloadInfo kOverloads[] = {
     /* num parameters */ 2,
     /* num template types */ 1,
     /* num template numbers */ 0,
-    /* template types */ &kTemplateTypes[28],
+    /* template types */ &kTemplateTypes[29],
     /* template numbers */ &kTemplateNumbers[10],
     /* parameters */ &kParameters[778],
     /* return matcher indices */ &kMatcherIndices[3],
@@ -13481,7 +13481,7 @@ constexpr OverloadInfo kOverloads[] = {
     /* num parameters */ 2,
     /* num template types */ 1,
     /* num template numbers */ 1,
-    /* template types */ &kTemplateTypes[28],
+    /* template types */ &kTemplateTypes[29],
     /* template numbers */ &kTemplateNumbers[4],
     /* parameters */ &kParameters[780],
     /* return matcher indices */ &kMatcherIndices[30],
@@ -14407,8 +14407,8 @@ constexpr IntrinsicInfo kBuiltins[] = {
   },
   {
     /* [69] */
-    /* fn sign<T : fa_f32_f16>(T) -> T */
+    /* fn sign<T : fia_fi32_f16>(T) -> T */
-    /* fn sign<N : num, T : fa_f32_f16>(vec<N, T>) -> vec<N, T> */
+    /* fn sign<N : num, T : fia_fi32_f16>(vec<N, T>) -> vec<N, T> */
     /* num overloads */ 2,
     /* overloads */ &kOverloads[387],
   },

src/tint/transform/builtin_polyfill.cc

@@ -590,6 +590,32 @@ struct BuiltinPolyfill::State {
         return name;
     }
 
+    /// Builds the polyfill function for the `sign` builtin when the element type is integer
+    /// @param ty the parameter and return type for the function
+    /// @return the polyfill function name
+    Symbol sign_int(const sem::Type* ty) {
+        const uint32_t width = WidthOf(ty);
+        auto zero = [&] { return ScalarOrVector(width, 0_a); };
+
+        // pos_or_neg_one = (v > 0) ? 1 : -1
+        auto pos_or_neg_one = b.Call("select",                     //
+                                     ScalarOrVector(width, -1_a),  //
+                                     ScalarOrVector(width, 1_a),   //
+                                     b.GreaterThan("v", zero()));
+
+        auto name = b.Symbols().New("tint_sign");
+        b.Func(name,
+               utils::Vector{
+                   b.Param("v", T(ty)),
+               },
+               T(ty),
+               utils::Vector{
+                   b.Return(b.Call("select", pos_or_neg_one, zero(), b.Equal("v", zero()))),
+               });
+
+        return name;
+    }
+
     /// Builds the polyfill function for the `textureSampleBaseClampToEdge` builtin, when the
     /// texture type is texture_2d<f32>.
     /// @return the polyfill function name
@@ -855,6 +881,15 @@ Transform::ApplyResult BuiltinPolyfill::Apply(const Program* src,
                             builtin, [&] { return s.saturate(builtin->ReturnType()); });
                     }
                     break;
+                case sem::BuiltinType::kSign:
+                    if (polyfill.sign_int) {
+                        auto* ty = builtin->ReturnType();
+                        if (ty->is_signed_integer_scalar_or_vector()) {
+                            fn = builtin_polyfills.GetOrCreate(builtin,
+                                                               [&] { return s.sign_int(ty); });
+                        }
+                    }
+                    break;
                 case sem::BuiltinType::kTextureSampleBaseClampToEdge:
                     if (polyfill.texture_sample_base_clamp_to_edge_2d_f32) {
                         auto& sig = builtin->Signature();

src/tint/transform/builtin_polyfill.h

@@ -68,6 +68,8 @@ class BuiltinPolyfill final : public Castable<BuiltinPolyfill, Transform> {
         bool int_div_mod = false;
         /// Should `saturate()` be polyfilled?
         bool saturate = false;
+        /// Should `sign()` be polyfilled for integer types?
+        bool sign_int = false;
         /// Should `textureSampleBaseClampToEdge()` be polyfilled for texture_2d<f32> textures?
         bool texture_sample_base_clamp_to_edge_2d_f32 = false;
         /// Should the vector form of `quantizeToF16()` be polyfilled with a scalar implementation?

src/tint/transform/builtin_polyfill_test.cc

@@ -34,11 +34,7 @@ TEST_F(BuiltinPolyfillTest, ShouldRunEmptyModule) {
 TEST_F(BuiltinPolyfillTest, EmptyModule) {
     auto* src = R"()";
 
-    auto* expect = src;
+    EXPECT_FALSE(ShouldRun<BuiltinPolyfill>(src));
-
-    auto got = Run<BuiltinPolyfill>(src);
-
-    EXPECT_EQ(expect, str(got));
 }
 
 ////////////////////////////////////////////////////////////////////////////////
@@ -73,11 +69,7 @@ fn f() {
 }
 )";
 
-    auto* expect = src;
+    EXPECT_FALSE(ShouldRun<BuiltinPolyfill>(src, polyfillAcosh(Level::kFull)));
-
-    auto got = Run<BuiltinPolyfill>(src, polyfillAcosh(Level::kFull));
-
-    EXPECT_EQ(expect, str(got));
 }
 
 TEST_F(BuiltinPolyfillTest, Acosh_Full_f32) {
@@ -206,11 +198,7 @@ fn f() {
 }
 )";
 
-    auto* expect = src;
+    EXPECT_FALSE(ShouldRun<BuiltinPolyfill>(src, polyfillSinh()));
-
-    auto got = Run<BuiltinPolyfill>(src, polyfillSinh());
-
-    EXPECT_EQ(expect, str(got));
 }
 
 TEST_F(BuiltinPolyfillTest, Asinh_f32) {
@@ -293,11 +281,7 @@ fn f() {
 }
 )";
 
-    auto* expect = src;
+    EXPECT_FALSE(ShouldRun<BuiltinPolyfill>(src, polyfillAtanh(Level::kFull)));
-
-    auto got = Run<BuiltinPolyfill>(src, polyfillAtanh(Level::kFull));
-
-    EXPECT_EQ(expect, str(got));
 }
 
 TEST_F(BuiltinPolyfillTest, Atanh_Full_f32) {
@@ -603,11 +587,7 @@ fn f() {
 }
 )";
 
-    auto* expect = src;
+    EXPECT_FALSE(ShouldRun<BuiltinPolyfill>(src, polyfillClampInteger()));
-
-    auto got = Run<BuiltinPolyfill>(src, polyfillClampInteger());
-
-    EXPECT_EQ(expect, str(got));
 }
 
 TEST_F(BuiltinPolyfillTest, ClampInteger_i32) {
@@ -732,11 +712,7 @@ fn f() {
 }
 )";
 
-    auto* expect = src;
+    EXPECT_FALSE(ShouldRun<BuiltinPolyfill>(src, polyfillCountLeadingZeros()));
-
-    auto got = Run<BuiltinPolyfill>(src, polyfillCountLeadingZeros());
-
-    EXPECT_EQ(expect, str(got));
 }
 
 TEST_F(BuiltinPolyfillTest, CountLeadingZeros_i32) {
@@ -909,11 +885,7 @@ fn f() {
 }
 )";
 
-    auto* expect = src;
+    EXPECT_FALSE(ShouldRun<BuiltinPolyfill>(src, polyfillCountTrailingZeros()));
-
-    auto got = Run<BuiltinPolyfill>(src, polyfillCountTrailingZeros());
-
-    EXPECT_EQ(expect, str(got));
 }
 
 TEST_F(BuiltinPolyfillTest, CountTrailingZeros_i32) {
@@ -1088,11 +1060,7 @@ fn f() {
 }
 )";
 
-    auto* expect = src;
+    EXPECT_FALSE(ShouldRun<BuiltinPolyfill>(src, polyfillExtractBits(Level::kFull)));
-
-    auto got = Run<BuiltinPolyfill>(src, polyfillExtractBits(Level::kFull));
-
-    EXPECT_EQ(expect, str(got));
 }
 
 TEST_F(BuiltinPolyfillTest, ExtractBits_Full_i32) {
@@ -1345,11 +1313,7 @@ fn f() {
 }
 )";
 
-    auto* expect = src;
+    EXPECT_FALSE(ShouldRun<BuiltinPolyfill>(src, polyfillFirstLeadingBit()));
-
-    auto got = Run<BuiltinPolyfill>(src, polyfillFirstLeadingBit());
-
-    EXPECT_EQ(expect, str(got));
 }
 
 TEST_F(BuiltinPolyfillTest, FirstLeadingBit_i32) {
@@ -1522,11 +1486,7 @@ fn f() {
 }
 )";
 
-    auto* expect = src;
+    EXPECT_FALSE(ShouldRun<BuiltinPolyfill>(src, polyfillFirstTrailingBit()));
-
-    auto got = Run<BuiltinPolyfill>(src, polyfillFirstTrailingBit());
-
-    EXPECT_EQ(expect, str(got));
 }
 
 TEST_F(BuiltinPolyfillTest, FirstTrailingBit_i32) {
@@ -1701,11 +1661,7 @@ fn f() {
 }
 )";
 
-    auto* expect = src;
+    EXPECT_FALSE(ShouldRun<BuiltinPolyfill>(src, polyfillInsertBits(Level::kFull)));
-
-    auto got = Run<BuiltinPolyfill>(src, polyfillInsertBits(Level::kFull));
-
-    EXPECT_EQ(expect, str(got));
 }
 
 TEST_F(BuiltinPolyfillTest, InsertBits_Full_i32) {
@@ -2715,11 +2671,7 @@ fn f() {
 }
 )";
 
-    auto* expect = src;
+    EXPECT_FALSE(ShouldRun<BuiltinPolyfill>(src, polyfillSaturate()));
-
-    auto got = Run<BuiltinPolyfill>(src, polyfillSaturate());
-
-    EXPECT_EQ(expect, str(got));
 }
 
 TEST_F(BuiltinPolyfillTest, Saturate_f32) {
@@ -2826,6 +2778,99 @@ fn f() {
     EXPECT_EQ(expect, str(got));
 }
 
+////////////////////////////////////////////////////////////////////////////////
+// sign_int
+////////////////////////////////////////////////////////////////////////////////
+DataMap polyfillSignInt() {
+    BuiltinPolyfill::Builtins builtins;
+    builtins.sign_int = true;
+    DataMap data;
+    data.Add<BuiltinPolyfill::Config>(builtins);
+    return data;
+}
+
+TEST_F(BuiltinPolyfillTest, ShouldRunSign_i32) {
+    auto* src = R"(
+fn f() {
+  let v = 1i;
+  sign(v);
+}
+)";
+
+    EXPECT_FALSE(ShouldRun<BuiltinPolyfill>(src));
+    EXPECT_TRUE(ShouldRun<BuiltinPolyfill>(src, polyfillSignInt()));
+}
+
+TEST_F(BuiltinPolyfillTest, ShouldRunSign_f32) {
+    auto* src = R"(
+fn f() {
+  let v = 1f;
+  sign(v);
+}
+)";
+
+    EXPECT_FALSE(ShouldRun<BuiltinPolyfill>(src));
+    EXPECT_FALSE(ShouldRun<BuiltinPolyfill>(src, polyfillSignInt()));
+}
+
+TEST_F(BuiltinPolyfillTest, SignInt_ConstantExpression) {
+    auto* src = R"(
+fn f() {
+  let r : i32 = sign(1i);
+}
+)";
+
+    EXPECT_FALSE(ShouldRun<BuiltinPolyfill>(src, polyfillSignInt()));
+}
+
+TEST_F(BuiltinPolyfillTest, SignInt_i32) {
+    auto* src = R"(
+fn f() {
+  let v = 1i;
+  let r : i32 = sign(v);
+}
+)";
+
+    auto* expect = R"(
+fn tint_sign(v : i32) -> i32 {
+  return select(select(-1, 1, (v > 0)), 0, (v == 0));
+}
+
+fn f() {
+  let v = 1i;
+  let r : i32 = tint_sign(v);
+}
+)";
+
+    auto got = Run<BuiltinPolyfill>(src, polyfillSignInt());
+
+    EXPECT_EQ(expect, str(got));
+}
+
+TEST_F(BuiltinPolyfillTest, SignInt_vec3_i32) {
+    auto* src = R"(
+fn f() {
+  let v = 1i;
+  let r : vec3<i32> = sign(vec3<i32>(v));
+}
+)";
+
+    auto* expect = R"(
+fn tint_sign(v : vec3<i32>) -> vec3<i32> {
+  return select(select(vec3(-1), vec3(1), (v > vec3(0))), vec3(0), (v == vec3(0)));
+}
+
+fn f() {
+  let v = 1i;
+  let r : vec3<i32> = tint_sign(vec3<i32>(v));
+}
+)";
+
+    auto got = Run<BuiltinPolyfill>(src, polyfillSignInt());
+
+    EXPECT_EQ(expect, str(got));
+}
+
 ////////////////////////////////////////////////////////////////////////////////
 // textureSampleBaseClampToEdge
 ////////////////////////////////////////////////////////////////////////////////

src/tint/writer/msl/generator_impl.cc

@@ -181,6 +181,7 @@ SanitizedResult Sanitize(const Program* in, const Options& options) {
         polyfills.first_trailing_bit = true;
         polyfills.insert_bits = transform::BuiltinPolyfill::Level::kClampParameters;
         polyfills.int_div_mod = true;
+        polyfills.sign_int = true;
         polyfills.texture_sample_base_clamp_to_edge_2d_f32 = true;
         data.Add<transform::BuiltinPolyfill::Config>(polyfills);
         manager.Add<transform::BuiltinPolyfill>();

src/tint/writer/spirv/builder.cc

@@ -201,7 +201,11 @@ uint32_t builtin_to_glsl_method(const sem::Builtin* builtin) {
         case BuiltinType::kRound:
             return GLSLstd450RoundEven;
         case BuiltinType::kSign:
+            if (builtin->ReturnType()->is_signed_integer_scalar_or_vector()) {
+                return GLSLstd450SSign;
+            } else {
                 return GLSLstd450FSign;
+            }
         case BuiltinType::kSin:
             return GLSLstd450Sin;
         case BuiltinType::kSinh:

test/tint/builtins/gen/literal/sign/3233fa.wgsl

@@ -0,0 +1,43 @@
+// Copyright 2022 The Tint Authors.
+//
+// Licensed under the Apache License, Version 2.0 (the "License");
+// you may not use this file except in compliance with the License.
+// You may obtain a copy of the License at
+//
+//     http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing, software
+// distributed under the License is distributed on an "AS IS" BASIS,
+// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+// See the License for the specific language governing permissions and
+// limitations under the License.
+
+////////////////////////////////////////////////////////////////////////////////
+// File generated by tools/src/cmd/gen
+// using the template:
+//   test/tint/builtins/gen/gen.wgsl.tmpl
+//
+// Do not modify this file directly
+////////////////////////////////////////////////////////////////////////////////
+
+
+// fn sign(i32) -> i32
+fn sign_3233fa() {
+  var res: i32 = sign(1i);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  sign_3233fa();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  sign_3233fa();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  sign_3233fa();
+}

test/tint/builtins/gen/literal/sign/3233fa.wgsl.expected.dxc.hlsl

@@ -0,0 +1,30 @@
+void sign_3233fa() {
+  int res = 1;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  sign_3233fa();
+  return (0.0f).xxxx;
+}
+
+tint_symbol vertex_main() {
+  const float4 inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = (tint_symbol)0;
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+void fragment_main() {
+  sign_3233fa();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  sign_3233fa();
+  return;
+}

test/tint/builtins/gen/literal/sign/3233fa.wgsl.expected.fxc.hlsl

@@ -0,0 +1,30 @@
+void sign_3233fa() {
+  int res = 1;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  sign_3233fa();
+  return (0.0f).xxxx;
+}
+
+tint_symbol vertex_main() {
+  const float4 inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = (tint_symbol)0;
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+void fragment_main() {
+  sign_3233fa();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  sign_3233fa();
+  return;
+}

test/tint/builtins/gen/literal/sign/3233fa.wgsl.expected.glsl

@@ -0,0 +1,49 @@
+#version 310 es
+
+void sign_3233fa() {
+  int res = 1;
+}
+
+vec4 vertex_main() {
+  sign_3233fa();
+  return vec4(0.0f);
+}
+
+void main() {
+  gl_PointSize = 1.0;
+  vec4 inner_result = vertex_main();
+  gl_Position = inner_result;
+  gl_Position.y = -(gl_Position.y);
+  gl_Position.z = ((2.0f * gl_Position.z) - gl_Position.w);
+  return;
+}
+#version 310 es
+precision mediump float;
+
+void sign_3233fa() {
+  int res = 1;
+}
+
+void fragment_main() {
+  sign_3233fa();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+void sign_3233fa() {
+  int res = 1;
+}
+
+void compute_main() {
+  sign_3233fa();
+}
+
+layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
+void main() {
+  compute_main();
+  return;
+}

test/tint/builtins/gen/literal/sign/3233fa.wgsl.expected.msl

@@ -0,0 +1,33 @@
+#include <metal_stdlib>
+
+using namespace metal;
+void sign_3233fa() {
+  int res = 1;
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  sign_3233fa();
+  return float4(0.0f);
+}
+
+vertex tint_symbol vertex_main() {
+  float4 const inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = {};
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+fragment void fragment_main() {
+  sign_3233fa();
+  return;
+}
+
+kernel void compute_main() {
+  sign_3233fa();
+  return;
+}
+

test/tint/builtins/gen/literal/sign/3233fa.wgsl.expected.spvasm

@@ -0,0 +1,66 @@
+; SPIR-V
+; Version: 1.3
+; Generator: Google Tint Compiler; 0
+; Bound: 32
+; Schema: 0
+               OpCapability Shader
+               OpMemoryModel Logical GLSL450
+               OpEntryPoint Vertex %vertex_main "vertex_main" %value %vertex_point_size
+               OpEntryPoint Fragment %fragment_main "fragment_main"
+               OpEntryPoint GLCompute %compute_main "compute_main"
+               OpExecutionMode %fragment_main OriginUpperLeft
+               OpExecutionMode %compute_main LocalSize 1 1 1
+               OpName %value "value"
+               OpName %vertex_point_size "vertex_point_size"
+               OpName %sign_3233fa "sign_3233fa"
+               OpName %res "res"
+               OpName %vertex_main_inner "vertex_main_inner"
+               OpName %vertex_main "vertex_main"
+               OpName %fragment_main "fragment_main"
+               OpName %compute_main "compute_main"
+               OpDecorate %value BuiltIn Position
+               OpDecorate %vertex_point_size BuiltIn PointSize
+      %float = OpTypeFloat 32
+    %v4float = OpTypeVector %float 4
+%_ptr_Output_v4float = OpTypePointer Output %v4float
+          %5 = OpConstantNull %v4float
+      %value = OpVariable %_ptr_Output_v4float Output %5
+%_ptr_Output_float = OpTypePointer Output %float
+          %8 = OpConstantNull %float
+%vertex_point_size = OpVariable %_ptr_Output_float Output %8
+       %void = OpTypeVoid
+          %9 = OpTypeFunction %void
+        %int = OpTypeInt 32 1
+      %int_1 = OpConstant %int 1
+%_ptr_Function_int = OpTypePointer Function %int
+         %17 = OpConstantNull %int
+         %18 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%sign_3233fa = OpFunction %void None %9
+         %12 = OpLabel
+        %res = OpVariable %_ptr_Function_int Function %17
+               OpStore %res %int_1
+               OpReturn
+               OpFunctionEnd
+%vertex_main_inner = OpFunction %v4float None %18
+         %20 = OpLabel
+         %21 = OpFunctionCall %void %sign_3233fa
+               OpReturnValue %5
+               OpFunctionEnd
+%vertex_main = OpFunction %void None %9
+         %23 = OpLabel
+         %24 = OpFunctionCall %v4float %vertex_main_inner
+               OpStore %value %24
+               OpStore %vertex_point_size %float_1
+               OpReturn
+               OpFunctionEnd
+%fragment_main = OpFunction %void None %9
+         %27 = OpLabel
+         %28 = OpFunctionCall %void %sign_3233fa
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %30 = OpLabel
+         %31 = OpFunctionCall %void %sign_3233fa
+               OpReturn
+               OpFunctionEnd

test/tint/builtins/gen/literal/sign/3233fa.wgsl.expected.wgsl

@@ -0,0 +1,19 @@
+fn sign_3233fa() {
+  var res : i32 = sign(1i);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  sign_3233fa();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  sign_3233fa();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  sign_3233fa();
+}

test/tint/builtins/gen/literal/sign/3a39ac.wgsl

@@ -0,0 +1,43 @@
+// Copyright 2022 The Tint Authors.
+//
+// Licensed under the Apache License, Version 2.0 (the "License");
+// you may not use this file except in compliance with the License.
+// You may obtain a copy of the License at
+//
+//     http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing, software
+// distributed under the License is distributed on an "AS IS" BASIS,
+// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+// See the License for the specific language governing permissions and
+// limitations under the License.
+
+////////////////////////////////////////////////////////////////////////////////
+// File generated by tools/src/cmd/gen
+// using the template:
+//   test/tint/builtins/gen/gen.wgsl.tmpl
+//
+// Do not modify this file directly
+////////////////////////////////////////////////////////////////////////////////
+
+
+// fn sign(ia) -> ia
+fn sign_3a39ac() {
+  var res = sign(1);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  sign_3a39ac();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  sign_3a39ac();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  sign_3a39ac();
+}

test/tint/builtins/gen/literal/sign/3a39ac.wgsl.expected.dxc.hlsl

@@ -0,0 +1,30 @@
+void sign_3a39ac() {
+  int res = 1;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  sign_3a39ac();
+  return (0.0f).xxxx;
+}
+
+tint_symbol vertex_main() {
+  const float4 inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = (tint_symbol)0;
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+void fragment_main() {
+  sign_3a39ac();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  sign_3a39ac();
+  return;
+}

test/tint/builtins/gen/literal/sign/3a39ac.wgsl.expected.fxc.hlsl

@@ -0,0 +1,30 @@
+void sign_3a39ac() {
+  int res = 1;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  sign_3a39ac();
+  return (0.0f).xxxx;
+}
+
+tint_symbol vertex_main() {
+  const float4 inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = (tint_symbol)0;
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+void fragment_main() {
+  sign_3a39ac();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  sign_3a39ac();
+  return;
+}

test/tint/builtins/gen/literal/sign/3a39ac.wgsl.expected.glsl

@@ -0,0 +1,49 @@
+#version 310 es
+
+void sign_3a39ac() {
+  int res = 1;
+}
+
+vec4 vertex_main() {
+  sign_3a39ac();
+  return vec4(0.0f);
+}
+
+void main() {
+  gl_PointSize = 1.0;
+  vec4 inner_result = vertex_main();
+  gl_Position = inner_result;
+  gl_Position.y = -(gl_Position.y);
+  gl_Position.z = ((2.0f * gl_Position.z) - gl_Position.w);
+  return;
+}
+#version 310 es
+precision mediump float;
+
+void sign_3a39ac() {
+  int res = 1;
+}
+
+void fragment_main() {
+  sign_3a39ac();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+void sign_3a39ac() {
+  int res = 1;
+}
+
+void compute_main() {
+  sign_3a39ac();
+}
+
+layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
+void main() {
+  compute_main();
+  return;
+}

test/tint/builtins/gen/literal/sign/3a39ac.wgsl.expected.msl

@@ -0,0 +1,33 @@
+#include <metal_stdlib>
+
+using namespace metal;
+void sign_3a39ac() {
+  int res = 1;
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  sign_3a39ac();
+  return float4(0.0f);
+}
+
+vertex tint_symbol vertex_main() {
+  float4 const inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = {};
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+fragment void fragment_main() {
+  sign_3a39ac();
+  return;
+}
+
+kernel void compute_main() {
+  sign_3a39ac();
+  return;
+}
+

test/tint/builtins/gen/literal/sign/3a39ac.wgsl.expected.spvasm

@@ -0,0 +1,66 @@
+; SPIR-V
+; Version: 1.3
+; Generator: Google Tint Compiler; 0
+; Bound: 32
+; Schema: 0
+               OpCapability Shader
+               OpMemoryModel Logical GLSL450
+               OpEntryPoint Vertex %vertex_main "vertex_main" %value %vertex_point_size
+               OpEntryPoint Fragment %fragment_main "fragment_main"
+               OpEntryPoint GLCompute %compute_main "compute_main"
+               OpExecutionMode %fragment_main OriginUpperLeft
+               OpExecutionMode %compute_main LocalSize 1 1 1
+               OpName %value "value"
+               OpName %vertex_point_size "vertex_point_size"
+               OpName %sign_3a39ac "sign_3a39ac"
+               OpName %res "res"
+               OpName %vertex_main_inner "vertex_main_inner"
+               OpName %vertex_main "vertex_main"
+               OpName %fragment_main "fragment_main"
+               OpName %compute_main "compute_main"
+               OpDecorate %value BuiltIn Position
+               OpDecorate %vertex_point_size BuiltIn PointSize
+      %float = OpTypeFloat 32
+    %v4float = OpTypeVector %float 4
+%_ptr_Output_v4float = OpTypePointer Output %v4float
+          %5 = OpConstantNull %v4float
+      %value = OpVariable %_ptr_Output_v4float Output %5
+%_ptr_Output_float = OpTypePointer Output %float
+          %8 = OpConstantNull %float
+%vertex_point_size = OpVariable %_ptr_Output_float Output %8
+       %void = OpTypeVoid
+          %9 = OpTypeFunction %void
+        %int = OpTypeInt 32 1
+      %int_1 = OpConstant %int 1
+%_ptr_Function_int = OpTypePointer Function %int
+         %17 = OpConstantNull %int
+         %18 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%sign_3a39ac = OpFunction %void None %9
+         %12 = OpLabel
+        %res = OpVariable %_ptr_Function_int Function %17
+               OpStore %res %int_1
+               OpReturn
+               OpFunctionEnd
+%vertex_main_inner = OpFunction %v4float None %18
+         %20 = OpLabel
+         %21 = OpFunctionCall %void %sign_3a39ac
+               OpReturnValue %5
+               OpFunctionEnd
+%vertex_main = OpFunction %void None %9
+         %23 = OpLabel
+         %24 = OpFunctionCall %v4float %vertex_main_inner
+               OpStore %value %24
+               OpStore %vertex_point_size %float_1
+               OpReturn
+               OpFunctionEnd
+%fragment_main = OpFunction %void None %9
+         %27 = OpLabel
+         %28 = OpFunctionCall %void %sign_3a39ac
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %30 = OpLabel
+         %31 = OpFunctionCall %void %sign_3a39ac
+               OpReturn
+               OpFunctionEnd

test/tint/builtins/gen/literal/sign/3a39ac.wgsl.expected.wgsl

@@ -0,0 +1,19 @@
+fn sign_3a39ac() {
+  var res = sign(1);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  sign_3a39ac();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  sign_3a39ac();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  sign_3a39ac();
+}

test/tint/builtins/gen/literal/sign/3bdab6.wgsl

@@ -0,0 +1,43 @@
+// Copyright 2022 The Tint Authors.
+//
+// Licensed under the Apache License, Version 2.0 (the "License");
+// you may not use this file except in compliance with the License.
+// You may obtain a copy of the License at
+//
+//     http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing, software
+// distributed under the License is distributed on an "AS IS" BASIS,
+// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+// See the License for the specific language governing permissions and
+// limitations under the License.
+
+////////////////////////////////////////////////////////////////////////////////
+// File generated by tools/src/cmd/gen
+// using the template:
+//   test/tint/builtins/gen/gen.wgsl.tmpl
+//
+// Do not modify this file directly
+////////////////////////////////////////////////////////////////////////////////
+
+
+// fn sign(vec<4, ia>) -> vec<4, ia>
+fn sign_3bdab6() {
+  var res = sign(vec4(1));
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  sign_3bdab6();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  sign_3bdab6();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  sign_3bdab6();
+}

test/tint/builtins/gen/literal/sign/3bdab6.wgsl.expected.dxc.hlsl

@@ -0,0 +1,30 @@
+void sign_3bdab6() {
+  int4 res = (1).xxxx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  sign_3bdab6();
+  return (0.0f).xxxx;
+}
+
+tint_symbol vertex_main() {
+  const float4 inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = (tint_symbol)0;
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+void fragment_main() {
+  sign_3bdab6();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  sign_3bdab6();
+  return;
+}

test/tint/builtins/gen/literal/sign/3bdab6.wgsl.expected.fxc.hlsl

@@ -0,0 +1,30 @@
+void sign_3bdab6() {
+  int4 res = (1).xxxx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  sign_3bdab6();
+  return (0.0f).xxxx;
+}
+
+tint_symbol vertex_main() {
+  const float4 inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = (tint_symbol)0;
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+void fragment_main() {
+  sign_3bdab6();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  sign_3bdab6();
+  return;
+}

test/tint/builtins/gen/literal/sign/3bdab6.wgsl.expected.glsl

@@ -0,0 +1,49 @@
+#version 310 es
+
+void sign_3bdab6() {
+  ivec4 res = ivec4(1);
+}
+
+vec4 vertex_main() {
+  sign_3bdab6();
+  return vec4(0.0f);
+}
+
+void main() {
+  gl_PointSize = 1.0;
+  vec4 inner_result = vertex_main();
+  gl_Position = inner_result;
+  gl_Position.y = -(gl_Position.y);
+  gl_Position.z = ((2.0f * gl_Position.z) - gl_Position.w);
+  return;
+}
+#version 310 es
+precision mediump float;
+
+void sign_3bdab6() {
+  ivec4 res = ivec4(1);
+}
+
+void fragment_main() {
+  sign_3bdab6();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+void sign_3bdab6() {
+  ivec4 res = ivec4(1);
+}
+
+void compute_main() {
+  sign_3bdab6();
+}
+
+layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
+void main() {
+  compute_main();
+  return;
+}

test/tint/builtins/gen/literal/sign/3bdab6.wgsl.expected.msl

@@ -0,0 +1,33 @@
+#include <metal_stdlib>
+
+using namespace metal;
+void sign_3bdab6() {
+  int4 res = int4(1);
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  sign_3bdab6();
+  return float4(0.0f);
+}
+
+vertex tint_symbol vertex_main() {
+  float4 const inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = {};
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+fragment void fragment_main() {
+  sign_3bdab6();
+  return;
+}
+
+kernel void compute_main() {
+  sign_3bdab6();
+  return;
+}
+

test/tint/builtins/gen/literal/sign/3bdab6.wgsl.expected.spvasm

@@ -0,0 +1,68 @@
+; SPIR-V
+; Version: 1.3
+; Generator: Google Tint Compiler; 0
+; Bound: 34
+; Schema: 0
+               OpCapability Shader
+               OpMemoryModel Logical GLSL450
+               OpEntryPoint Vertex %vertex_main "vertex_main" %value %vertex_point_size
+               OpEntryPoint Fragment %fragment_main "fragment_main"
+               OpEntryPoint GLCompute %compute_main "compute_main"
+               OpExecutionMode %fragment_main OriginUpperLeft
+               OpExecutionMode %compute_main LocalSize 1 1 1
+               OpName %value "value"
+               OpName %vertex_point_size "vertex_point_size"
+               OpName %sign_3bdab6 "sign_3bdab6"
+               OpName %res "res"
+               OpName %vertex_main_inner "vertex_main_inner"
+               OpName %vertex_main "vertex_main"
+               OpName %fragment_main "fragment_main"
+               OpName %compute_main "compute_main"
+               OpDecorate %value BuiltIn Position
+               OpDecorate %vertex_point_size BuiltIn PointSize
+      %float = OpTypeFloat 32
+    %v4float = OpTypeVector %float 4
+%_ptr_Output_v4float = OpTypePointer Output %v4float
+          %5 = OpConstantNull %v4float
+      %value = OpVariable %_ptr_Output_v4float Output %5
+%_ptr_Output_float = OpTypePointer Output %float
+          %8 = OpConstantNull %float
+%vertex_point_size = OpVariable %_ptr_Output_float Output %8
+       %void = OpTypeVoid
+          %9 = OpTypeFunction %void
+        %int = OpTypeInt 32 1
+      %v4int = OpTypeVector %int 4
+      %int_1 = OpConstant %int 1
+         %16 = OpConstantComposite %v4int %int_1 %int_1 %int_1 %int_1
+%_ptr_Function_v4int = OpTypePointer Function %v4int
+         %19 = OpConstantNull %v4int
+         %20 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%sign_3bdab6 = OpFunction %void None %9
+         %12 = OpLabel
+        %res = OpVariable %_ptr_Function_v4int Function %19
+               OpStore %res %16
+               OpReturn
+               OpFunctionEnd
+%vertex_main_inner = OpFunction %v4float None %20
+         %22 = OpLabel
+         %23 = OpFunctionCall %void %sign_3bdab6
+               OpReturnValue %5
+               OpFunctionEnd
+%vertex_main = OpFunction %void None %9
+         %25 = OpLabel
+         %26 = OpFunctionCall %v4float %vertex_main_inner
+               OpStore %value %26
+               OpStore %vertex_point_size %float_1
+               OpReturn
+               OpFunctionEnd
+%fragment_main = OpFunction %void None %9
+         %29 = OpLabel
+         %30 = OpFunctionCall %void %sign_3bdab6
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %32 = OpLabel
+         %33 = OpFunctionCall %void %sign_3bdab6
+               OpReturn
+               OpFunctionEnd

test/tint/builtins/gen/literal/sign/3bdab6.wgsl.expected.wgsl

@@ -0,0 +1,19 @@
+fn sign_3bdab6() {
+  var res = sign(vec4(1));
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  sign_3bdab6();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  sign_3bdab6();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  sign_3bdab6();
+}

test/tint/builtins/gen/literal/sign/55339e.wgsl

@@ -0,0 +1,43 @@
+// Copyright 2022 The Tint Authors.
+//
+// Licensed under the Apache License, Version 2.0 (the "License");
+// you may not use this file except in compliance with the License.
+// You may obtain a copy of the License at
+//
+//     http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing, software
+// distributed under the License is distributed on an "AS IS" BASIS,
+// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+// See the License for the specific language governing permissions and
+// limitations under the License.
+
+////////////////////////////////////////////////////////////////////////////////
+// File generated by tools/src/cmd/gen
+// using the template:
+//   test/tint/builtins/gen/gen.wgsl.tmpl
+//
+// Do not modify this file directly
+////////////////////////////////////////////////////////////////////////////////
+
+
+// fn sign(vec<3, ia>) -> vec<3, ia>
+fn sign_55339e() {
+  var res = sign(vec3(1));
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  sign_55339e();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  sign_55339e();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  sign_55339e();
+}

test/tint/builtins/gen/literal/sign/55339e.wgsl.expected.dxc.hlsl

@@ -0,0 +1,30 @@
+void sign_55339e() {
+  int3 res = (1).xxx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  sign_55339e();
+  return (0.0f).xxxx;
+}
+
+tint_symbol vertex_main() {
+  const float4 inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = (tint_symbol)0;
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+void fragment_main() {
+  sign_55339e();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  sign_55339e();
+  return;
+}

test/tint/builtins/gen/literal/sign/55339e.wgsl.expected.fxc.hlsl

@@ -0,0 +1,30 @@
+void sign_55339e() {
+  int3 res = (1).xxx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  sign_55339e();
+  return (0.0f).xxxx;
+}
+
+tint_symbol vertex_main() {
+  const float4 inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = (tint_symbol)0;
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+void fragment_main() {
+  sign_55339e();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  sign_55339e();
+  return;
+}

test/tint/builtins/gen/literal/sign/55339e.wgsl.expected.glsl

@@ -0,0 +1,49 @@
+#version 310 es
+
+void sign_55339e() {
+  ivec3 res = ivec3(1);
+}
+
+vec4 vertex_main() {
+  sign_55339e();
+  return vec4(0.0f);
+}
+
+void main() {
+  gl_PointSize = 1.0;
+  vec4 inner_result = vertex_main();
+  gl_Position = inner_result;
+  gl_Position.y = -(gl_Position.y);
+  gl_Position.z = ((2.0f * gl_Position.z) - gl_Position.w);
+  return;
+}
+#version 310 es
+precision mediump float;
+
+void sign_55339e() {
+  ivec3 res = ivec3(1);
+}
+
+void fragment_main() {
+  sign_55339e();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+void sign_55339e() {
+  ivec3 res = ivec3(1);
+}
+
+void compute_main() {
+  sign_55339e();
+}
+
+layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
+void main() {
+  compute_main();
+  return;
+}

test/tint/builtins/gen/literal/sign/55339e.wgsl.expected.msl

@@ -0,0 +1,33 @@
+#include <metal_stdlib>
+
+using namespace metal;
+void sign_55339e() {
+  int3 res = int3(1);
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  sign_55339e();
+  return float4(0.0f);
+}
+
+vertex tint_symbol vertex_main() {
+  float4 const inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = {};
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+fragment void fragment_main() {
+  sign_55339e();
+  return;
+}
+
+kernel void compute_main() {
+  sign_55339e();
+  return;
+}
+

test/tint/builtins/gen/literal/sign/55339e.wgsl.expected.spvasm

@@ -0,0 +1,68 @@
+; SPIR-V
+; Version: 1.3
+; Generator: Google Tint Compiler; 0
+; Bound: 34
+; Schema: 0
+               OpCapability Shader
+               OpMemoryModel Logical GLSL450
+               OpEntryPoint Vertex %vertex_main "vertex_main" %value %vertex_point_size
+               OpEntryPoint Fragment %fragment_main "fragment_main"
+               OpEntryPoint GLCompute %compute_main "compute_main"
+               OpExecutionMode %fragment_main OriginUpperLeft
+               OpExecutionMode %compute_main LocalSize 1 1 1
+               OpName %value "value"
+               OpName %vertex_point_size "vertex_point_size"
+               OpName %sign_55339e "sign_55339e"
+               OpName %res "res"
+               OpName %vertex_main_inner "vertex_main_inner"
+               OpName %vertex_main "vertex_main"
+               OpName %fragment_main "fragment_main"
+               OpName %compute_main "compute_main"
+               OpDecorate %value BuiltIn Position
+               OpDecorate %vertex_point_size BuiltIn PointSize
+      %float = OpTypeFloat 32
+    %v4float = OpTypeVector %float 4
+%_ptr_Output_v4float = OpTypePointer Output %v4float
+          %5 = OpConstantNull %v4float
+      %value = OpVariable %_ptr_Output_v4float Output %5
+%_ptr_Output_float = OpTypePointer Output %float
+          %8 = OpConstantNull %float
+%vertex_point_size = OpVariable %_ptr_Output_float Output %8
+       %void = OpTypeVoid
+          %9 = OpTypeFunction %void
+        %int = OpTypeInt 32 1
+      %v3int = OpTypeVector %int 3
+      %int_1 = OpConstant %int 1
+         %16 = OpConstantComposite %v3int %int_1 %int_1 %int_1
+%_ptr_Function_v3int = OpTypePointer Function %v3int
+         %19 = OpConstantNull %v3int
+         %20 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%sign_55339e = OpFunction %void None %9
+         %12 = OpLabel
+        %res = OpVariable %_ptr_Function_v3int Function %19
+               OpStore %res %16
+               OpReturn
+               OpFunctionEnd
+%vertex_main_inner = OpFunction %v4float None %20
+         %22 = OpLabel
+         %23 = OpFunctionCall %void %sign_55339e
+               OpReturnValue %5
+               OpFunctionEnd
+%vertex_main = OpFunction %void None %9
+         %25 = OpLabel
+         %26 = OpFunctionCall %v4float %vertex_main_inner
+               OpStore %value %26
+               OpStore %vertex_point_size %float_1
+               OpReturn
+               OpFunctionEnd
+%fragment_main = OpFunction %void None %9
+         %29 = OpLabel
+         %30 = OpFunctionCall %void %sign_55339e
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %32 = OpLabel
+         %33 = OpFunctionCall %void %sign_55339e
+               OpReturn
+               OpFunctionEnd

test/tint/builtins/gen/literal/sign/55339e.wgsl.expected.wgsl

@@ -0,0 +1,19 @@
+fn sign_55339e() {
+  var res = sign(vec3(1));
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  sign_55339e();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  sign_55339e();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  sign_55339e();
+}

test/tint/builtins/gen/literal/sign/58d779.wgsl

@@ -0,0 +1,43 @@
+// Copyright 2022 The Tint Authors.
+//
+// Licensed under the Apache License, Version 2.0 (the "License");
+// you may not use this file except in compliance with the License.
+// You may obtain a copy of the License at
+//
+//     http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing, software
+// distributed under the License is distributed on an "AS IS" BASIS,
+// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+// See the License for the specific language governing permissions and
+// limitations under the License.
+
+////////////////////////////////////////////////////////////////////////////////
+// File generated by tools/src/cmd/gen
+// using the template:
+//   test/tint/builtins/gen/gen.wgsl.tmpl
+//
+// Do not modify this file directly
+////////////////////////////////////////////////////////////////////////////////
+
+
+// fn sign(vec<4, i32>) -> vec<4, i32>
+fn sign_58d779() {
+  var res: vec4<i32> = sign(vec4<i32>(1i));
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  sign_58d779();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  sign_58d779();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  sign_58d779();
+}

test/tint/builtins/gen/literal/sign/58d779.wgsl.expected.dxc.hlsl

@@ -0,0 +1,30 @@
+void sign_58d779() {
+  int4 res = (1).xxxx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  sign_58d779();
+  return (0.0f).xxxx;
+}
+
+tint_symbol vertex_main() {
+  const float4 inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = (tint_symbol)0;
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+void fragment_main() {
+  sign_58d779();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  sign_58d779();
+  return;
+}

test/tint/builtins/gen/literal/sign/58d779.wgsl.expected.fxc.hlsl

@@ -0,0 +1,30 @@
+void sign_58d779() {
+  int4 res = (1).xxxx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  sign_58d779();
+  return (0.0f).xxxx;
+}
+
+tint_symbol vertex_main() {
+  const float4 inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = (tint_symbol)0;
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+void fragment_main() {
+  sign_58d779();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  sign_58d779();
+  return;
+}

test/tint/builtins/gen/literal/sign/58d779.wgsl.expected.glsl

@@ -0,0 +1,49 @@
+#version 310 es
+
+void sign_58d779() {
+  ivec4 res = ivec4(1);
+}
+
+vec4 vertex_main() {
+  sign_58d779();
+  return vec4(0.0f);
+}
+
+void main() {
+  gl_PointSize = 1.0;
+  vec4 inner_result = vertex_main();
+  gl_Position = inner_result;
+  gl_Position.y = -(gl_Position.y);
+  gl_Position.z = ((2.0f * gl_Position.z) - gl_Position.w);
+  return;
+}
+#version 310 es
+precision mediump float;
+
+void sign_58d779() {
+  ivec4 res = ivec4(1);
+}
+
+void fragment_main() {
+  sign_58d779();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+void sign_58d779() {
+  ivec4 res = ivec4(1);
+}
+
+void compute_main() {
+  sign_58d779();
+}
+
+layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
+void main() {
+  compute_main();
+  return;
+}

test/tint/builtins/gen/literal/sign/58d779.wgsl.expected.msl

@@ -0,0 +1,33 @@
+#include <metal_stdlib>
+
+using namespace metal;
+void sign_58d779() {
+  int4 res = int4(1);
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  sign_58d779();
+  return float4(0.0f);
+}
+
+vertex tint_symbol vertex_main() {
+  float4 const inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = {};
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+fragment void fragment_main() {
+  sign_58d779();
+  return;
+}
+
+kernel void compute_main() {
+  sign_58d779();
+  return;
+}
+

test/tint/builtins/gen/literal/sign/58d779.wgsl.expected.spvasm

@@ -0,0 +1,68 @@
+; SPIR-V
+; Version: 1.3
+; Generator: Google Tint Compiler; 0
+; Bound: 34
+; Schema: 0
+               OpCapability Shader
+               OpMemoryModel Logical GLSL450
+               OpEntryPoint Vertex %vertex_main "vertex_main" %value %vertex_point_size
+               OpEntryPoint Fragment %fragment_main "fragment_main"
+               OpEntryPoint GLCompute %compute_main "compute_main"
+               OpExecutionMode %fragment_main OriginUpperLeft
+               OpExecutionMode %compute_main LocalSize 1 1 1
+               OpName %value "value"
+               OpName %vertex_point_size "vertex_point_size"
+               OpName %sign_58d779 "sign_58d779"
+               OpName %res "res"
+               OpName %vertex_main_inner "vertex_main_inner"
+               OpName %vertex_main "vertex_main"
+               OpName %fragment_main "fragment_main"
+               OpName %compute_main "compute_main"
+               OpDecorate %value BuiltIn Position
+               OpDecorate %vertex_point_size BuiltIn PointSize
+      %float = OpTypeFloat 32
+    %v4float = OpTypeVector %float 4
+%_ptr_Output_v4float = OpTypePointer Output %v4float
+          %5 = OpConstantNull %v4float
+      %value = OpVariable %_ptr_Output_v4float Output %5
+%_ptr_Output_float = OpTypePointer Output %float
+          %8 = OpConstantNull %float
+%vertex_point_size = OpVariable %_ptr_Output_float Output %8
+       %void = OpTypeVoid
+          %9 = OpTypeFunction %void
+        %int = OpTypeInt 32 1
+      %v4int = OpTypeVector %int 4
+      %int_1 = OpConstant %int 1
+         %16 = OpConstantComposite %v4int %int_1 %int_1 %int_1 %int_1
+%_ptr_Function_v4int = OpTypePointer Function %v4int
+         %19 = OpConstantNull %v4int
+         %20 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%sign_58d779 = OpFunction %void None %9
+         %12 = OpLabel
+        %res = OpVariable %_ptr_Function_v4int Function %19
+               OpStore %res %16
+               OpReturn
+               OpFunctionEnd
+%vertex_main_inner = OpFunction %v4float None %20
+         %22 = OpLabel
+         %23 = OpFunctionCall %void %sign_58d779
+               OpReturnValue %5
+               OpFunctionEnd
+%vertex_main = OpFunction %void None %9
+         %25 = OpLabel
+         %26 = OpFunctionCall %v4float %vertex_main_inner
+               OpStore %value %26
+               OpStore %vertex_point_size %float_1
+               OpReturn
+               OpFunctionEnd
+%fragment_main = OpFunction %void None %9
+         %29 = OpLabel
+         %30 = OpFunctionCall %void %sign_58d779
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %32 = OpLabel
+         %33 = OpFunctionCall %void %sign_58d779
+               OpReturn
+               OpFunctionEnd

test/tint/builtins/gen/literal/sign/58d779.wgsl.expected.wgsl

@@ -0,0 +1,19 @@
+fn sign_58d779() {
+  var res : vec4<i32> = sign(vec4<i32>(1i));
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  sign_58d779();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  sign_58d779();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  sign_58d779();
+}

test/tint/builtins/gen/literal/sign/926015.wgsl

@@ -0,0 +1,43 @@
+// Copyright 2022 The Tint Authors.
+//
+// Licensed under the Apache License, Version 2.0 (the "License");
+// you may not use this file except in compliance with the License.
+// You may obtain a copy of the License at
+//
+//     http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing, software
+// distributed under the License is distributed on an "AS IS" BASIS,
+// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+// See the License for the specific language governing permissions and
+// limitations under the License.
+
+////////////////////////////////////////////////////////////////////////////////
+// File generated by tools/src/cmd/gen
+// using the template:
+//   test/tint/builtins/gen/gen.wgsl.tmpl
+//
+// Do not modify this file directly
+////////////////////////////////////////////////////////////////////////////////
+
+
+// fn sign(vec<2, i32>) -> vec<2, i32>
+fn sign_926015() {
+  var res: vec2<i32> = sign(vec2<i32>(1i));
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  sign_926015();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  sign_926015();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  sign_926015();
+}

test/tint/builtins/gen/literal/sign/926015.wgsl.expected.dxc.hlsl

@@ -0,0 +1,30 @@
+void sign_926015() {
+  int2 res = (1).xx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  sign_926015();
+  return (0.0f).xxxx;
+}
+
+tint_symbol vertex_main() {
+  const float4 inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = (tint_symbol)0;
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+void fragment_main() {
+  sign_926015();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  sign_926015();
+  return;
+}

test/tint/builtins/gen/literal/sign/926015.wgsl.expected.fxc.hlsl

@@ -0,0 +1,30 @@
+void sign_926015() {
+  int2 res = (1).xx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  sign_926015();
+  return (0.0f).xxxx;
+}
+
+tint_symbol vertex_main() {
+  const float4 inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = (tint_symbol)0;
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+void fragment_main() {
+  sign_926015();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  sign_926015();
+  return;
+}

test/tint/builtins/gen/literal/sign/926015.wgsl.expected.glsl

@@ -0,0 +1,49 @@
+#version 310 es
+
+void sign_926015() {
+  ivec2 res = ivec2(1);
+}
+
+vec4 vertex_main() {
+  sign_926015();
+  return vec4(0.0f);
+}
+
+void main() {
+  gl_PointSize = 1.0;
+  vec4 inner_result = vertex_main();
+  gl_Position = inner_result;
+  gl_Position.y = -(gl_Position.y);
+  gl_Position.z = ((2.0f * gl_Position.z) - gl_Position.w);
+  return;
+}
+#version 310 es
+precision mediump float;
+
+void sign_926015() {
+  ivec2 res = ivec2(1);
+}
+
+void fragment_main() {
+  sign_926015();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+void sign_926015() {
+  ivec2 res = ivec2(1);
+}
+
+void compute_main() {
+  sign_926015();
+}
+
+layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
+void main() {
+  compute_main();
+  return;
+}

test/tint/builtins/gen/literal/sign/926015.wgsl.expected.msl

@@ -0,0 +1,33 @@
+#include <metal_stdlib>
+
+using namespace metal;
+void sign_926015() {
+  int2 res = int2(1);
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  sign_926015();
+  return float4(0.0f);
+}
+
+vertex tint_symbol vertex_main() {
+  float4 const inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = {};
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+fragment void fragment_main() {
+  sign_926015();
+  return;
+}
+
+kernel void compute_main() {
+  sign_926015();
+  return;
+}
+

test/tint/builtins/gen/literal/sign/926015.wgsl.expected.spvasm

@@ -0,0 +1,68 @@
+; SPIR-V
+; Version: 1.3
+; Generator: Google Tint Compiler; 0
+; Bound: 34
+; Schema: 0
+               OpCapability Shader
+               OpMemoryModel Logical GLSL450
+               OpEntryPoint Vertex %vertex_main "vertex_main" %value %vertex_point_size
+               OpEntryPoint Fragment %fragment_main "fragment_main"
+               OpEntryPoint GLCompute %compute_main "compute_main"
+               OpExecutionMode %fragment_main OriginUpperLeft
+               OpExecutionMode %compute_main LocalSize 1 1 1
+               OpName %value "value"
+               OpName %vertex_point_size "vertex_point_size"
+               OpName %sign_926015 "sign_926015"
+               OpName %res "res"
+               OpName %vertex_main_inner "vertex_main_inner"
+               OpName %vertex_main "vertex_main"
+               OpName %fragment_main "fragment_main"
+               OpName %compute_main "compute_main"
+               OpDecorate %value BuiltIn Position
+               OpDecorate %vertex_point_size BuiltIn PointSize
+      %float = OpTypeFloat 32
+    %v4float = OpTypeVector %float 4
+%_ptr_Output_v4float = OpTypePointer Output %v4float
+          %5 = OpConstantNull %v4float
+      %value = OpVariable %_ptr_Output_v4float Output %5
+%_ptr_Output_float = OpTypePointer Output %float
+          %8 = OpConstantNull %float
+%vertex_point_size = OpVariable %_ptr_Output_float Output %8
+       %void = OpTypeVoid
+          %9 = OpTypeFunction %void
+        %int = OpTypeInt 32 1
+      %v2int = OpTypeVector %int 2
+      %int_1 = OpConstant %int 1
+         %16 = OpConstantComposite %v2int %int_1 %int_1
+%_ptr_Function_v2int = OpTypePointer Function %v2int
+         %19 = OpConstantNull %v2int
+         %20 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%sign_926015 = OpFunction %void None %9
+         %12 = OpLabel
+        %res = OpVariable %_ptr_Function_v2int Function %19
+               OpStore %res %16
+               OpReturn
+               OpFunctionEnd
+%vertex_main_inner = OpFunction %v4float None %20
+         %22 = OpLabel
+         %23 = OpFunctionCall %void %sign_926015
+               OpReturnValue %5
+               OpFunctionEnd
+%vertex_main = OpFunction %void None %9
+         %25 = OpLabel
+         %26 = OpFunctionCall %v4float %vertex_main_inner
+               OpStore %value %26
+               OpStore %vertex_point_size %float_1
+               OpReturn
+               OpFunctionEnd
+%fragment_main = OpFunction %void None %9
+         %29 = OpLabel
+         %30 = OpFunctionCall %void %sign_926015
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %32 = OpLabel
+         %33 = OpFunctionCall %void %sign_926015
+               OpReturn
+               OpFunctionEnd

test/tint/builtins/gen/literal/sign/926015.wgsl.expected.wgsl

@@ -0,0 +1,19 @@
+fn sign_926015() {
+  var res : vec2<i32> = sign(vec2<i32>(1i));
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  sign_926015();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  sign_926015();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  sign_926015();
+}

test/tint/builtins/gen/literal/sign/943b2e.wgsl

@@ -0,0 +1,43 @@
+// Copyright 2022 The Tint Authors.
+//
+// Licensed under the Apache License, Version 2.0 (the "License");
+// you may not use this file except in compliance with the License.
+// You may obtain a copy of the License at
+//
+//     http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing, software
+// distributed under the License is distributed on an "AS IS" BASIS,
+// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+// See the License for the specific language governing permissions and
+// limitations under the License.
+
+////////////////////////////////////////////////////////////////////////////////
+// File generated by tools/src/cmd/gen
+// using the template:
+//   test/tint/builtins/gen/gen.wgsl.tmpl
+//
+// Do not modify this file directly
+////////////////////////////////////////////////////////////////////////////////
+
+
+// fn sign(vec<2, ia>) -> vec<2, ia>
+fn sign_943b2e() {
+  var res = sign(vec2(1));
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  sign_943b2e();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  sign_943b2e();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  sign_943b2e();
+}

test/tint/builtins/gen/literal/sign/943b2e.wgsl.expected.dxc.hlsl

@@ -0,0 +1,30 @@
+void sign_943b2e() {
+  int2 res = (1).xx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  sign_943b2e();
+  return (0.0f).xxxx;
+}
+
+tint_symbol vertex_main() {
+  const float4 inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = (tint_symbol)0;
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+void fragment_main() {
+  sign_943b2e();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  sign_943b2e();
+  return;
+}

test/tint/builtins/gen/literal/sign/943b2e.wgsl.expected.fxc.hlsl

@@ -0,0 +1,30 @@
+void sign_943b2e() {
+  int2 res = (1).xx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  sign_943b2e();
+  return (0.0f).xxxx;
+}
+
+tint_symbol vertex_main() {
+  const float4 inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = (tint_symbol)0;
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+void fragment_main() {
+  sign_943b2e();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  sign_943b2e();
+  return;
+}

test/tint/builtins/gen/literal/sign/943b2e.wgsl.expected.glsl

@@ -0,0 +1,49 @@
+#version 310 es
+
+void sign_943b2e() {
+  ivec2 res = ivec2(1);
+}
+
+vec4 vertex_main() {
+  sign_943b2e();
+  return vec4(0.0f);
+}
+
+void main() {
+  gl_PointSize = 1.0;
+  vec4 inner_result = vertex_main();
+  gl_Position = inner_result;
+  gl_Position.y = -(gl_Position.y);
+  gl_Position.z = ((2.0f * gl_Position.z) - gl_Position.w);
+  return;
+}
+#version 310 es
+precision mediump float;
+
+void sign_943b2e() {
+  ivec2 res = ivec2(1);
+}
+
+void fragment_main() {
+  sign_943b2e();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+void sign_943b2e() {
+  ivec2 res = ivec2(1);
+}
+
+void compute_main() {
+  sign_943b2e();
+}
+
+layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
+void main() {
+  compute_main();
+  return;
+}

test/tint/builtins/gen/literal/sign/943b2e.wgsl.expected.msl

@@ -0,0 +1,33 @@
+#include <metal_stdlib>
+
+using namespace metal;
+void sign_943b2e() {
+  int2 res = int2(1);
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  sign_943b2e();
+  return float4(0.0f);
+}
+
+vertex tint_symbol vertex_main() {
+  float4 const inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = {};
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+fragment void fragment_main() {
+  sign_943b2e();
+  return;
+}
+
+kernel void compute_main() {
+  sign_943b2e();
+  return;
+}
+

test/tint/builtins/gen/literal/sign/943b2e.wgsl.expected.spvasm

@@ -0,0 +1,68 @@
+; SPIR-V
+; Version: 1.3
+; Generator: Google Tint Compiler; 0
+; Bound: 34
+; Schema: 0
+               OpCapability Shader
+               OpMemoryModel Logical GLSL450
+               OpEntryPoint Vertex %vertex_main "vertex_main" %value %vertex_point_size
+               OpEntryPoint Fragment %fragment_main "fragment_main"
+               OpEntryPoint GLCompute %compute_main "compute_main"
+               OpExecutionMode %fragment_main OriginUpperLeft
+               OpExecutionMode %compute_main LocalSize 1 1 1
+               OpName %value "value"
+               OpName %vertex_point_size "vertex_point_size"
+               OpName %sign_943b2e "sign_943b2e"
+               OpName %res "res"
+               OpName %vertex_main_inner "vertex_main_inner"
+               OpName %vertex_main "vertex_main"
+               OpName %fragment_main "fragment_main"
+               OpName %compute_main "compute_main"
+               OpDecorate %value BuiltIn Position
+               OpDecorate %vertex_point_size BuiltIn PointSize
+      %float = OpTypeFloat 32
+    %v4float = OpTypeVector %float 4
+%_ptr_Output_v4float = OpTypePointer Output %v4float
+          %5 = OpConstantNull %v4float
+      %value = OpVariable %_ptr_Output_v4float Output %5
+%_ptr_Output_float = OpTypePointer Output %float
+          %8 = OpConstantNull %float
+%vertex_point_size = OpVariable %_ptr_Output_float Output %8
+       %void = OpTypeVoid
+          %9 = OpTypeFunction %void
+        %int = OpTypeInt 32 1
+      %v2int = OpTypeVector %int 2
+      %int_1 = OpConstant %int 1
+         %16 = OpConstantComposite %v2int %int_1 %int_1
+%_ptr_Function_v2int = OpTypePointer Function %v2int
+         %19 = OpConstantNull %v2int
+         %20 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%sign_943b2e = OpFunction %void None %9
+         %12 = OpLabel
+        %res = OpVariable %_ptr_Function_v2int Function %19
+               OpStore %res %16
+               OpReturn
+               OpFunctionEnd
+%vertex_main_inner = OpFunction %v4float None %20
+         %22 = OpLabel
+         %23 = OpFunctionCall %void %sign_943b2e
+               OpReturnValue %5
+               OpFunctionEnd
+%vertex_main = OpFunction %void None %9
+         %25 = OpLabel
+         %26 = OpFunctionCall %v4float %vertex_main_inner
+               OpStore %value %26
+               OpStore %vertex_point_size %float_1
+               OpReturn
+               OpFunctionEnd
+%fragment_main = OpFunction %void None %9
+         %29 = OpLabel
+         %30 = OpFunctionCall %void %sign_943b2e
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %32 = OpLabel
+         %33 = OpFunctionCall %void %sign_943b2e
+               OpReturn
+               OpFunctionEnd

test/tint/builtins/gen/literal/sign/943b2e.wgsl.expected.wgsl

@@ -0,0 +1,19 @@
+fn sign_943b2e() {
+  var res = sign(vec2(1));
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  sign_943b2e();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  sign_943b2e();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  sign_943b2e();
+}

test/tint/builtins/gen/literal/sign/9603b1.wgsl

@@ -0,0 +1,43 @@
+// Copyright 2022 The Tint Authors.
+//
+// Licensed under the Apache License, Version 2.0 (the "License");
+// you may not use this file except in compliance with the License.
+// You may obtain a copy of the License at
+//
+//     http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing, software
+// distributed under the License is distributed on an "AS IS" BASIS,
+// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+// See the License for the specific language governing permissions and
+// limitations under the License.
+
+////////////////////////////////////////////////////////////////////////////////
+// File generated by tools/src/cmd/gen
+// using the template:
+//   test/tint/builtins/gen/gen.wgsl.tmpl
+//
+// Do not modify this file directly
+////////////////////////////////////////////////////////////////////////////////
+
+
+// fn sign(vec<3, i32>) -> vec<3, i32>
+fn sign_9603b1() {
+  var res: vec3<i32> = sign(vec3<i32>(1i));
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  sign_9603b1();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  sign_9603b1();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  sign_9603b1();
+}

test/tint/builtins/gen/literal/sign/9603b1.wgsl.expected.dxc.hlsl

@@ -0,0 +1,30 @@
+void sign_9603b1() {
+  int3 res = (1).xxx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  sign_9603b1();
+  return (0.0f).xxxx;
+}
+
+tint_symbol vertex_main() {
+  const float4 inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = (tint_symbol)0;
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+void fragment_main() {
+  sign_9603b1();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  sign_9603b1();
+  return;
+}

test/tint/builtins/gen/literal/sign/9603b1.wgsl.expected.fxc.hlsl

@@ -0,0 +1,30 @@
+void sign_9603b1() {
+  int3 res = (1).xxx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  sign_9603b1();
+  return (0.0f).xxxx;
+}
+
+tint_symbol vertex_main() {
+  const float4 inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = (tint_symbol)0;
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+void fragment_main() {
+  sign_9603b1();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  sign_9603b1();
+  return;
+}

test/tint/builtins/gen/literal/sign/9603b1.wgsl.expected.glsl

@@ -0,0 +1,49 @@
+#version 310 es
+
+void sign_9603b1() {
+  ivec3 res = ivec3(1);
+}
+
+vec4 vertex_main() {
+  sign_9603b1();
+  return vec4(0.0f);
+}
+
+void main() {
+  gl_PointSize = 1.0;
+  vec4 inner_result = vertex_main();
+  gl_Position = inner_result;
+  gl_Position.y = -(gl_Position.y);
+  gl_Position.z = ((2.0f * gl_Position.z) - gl_Position.w);
+  return;
+}
+#version 310 es
+precision mediump float;
+
+void sign_9603b1() {
+  ivec3 res = ivec3(1);
+}
+
+void fragment_main() {
+  sign_9603b1();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+void sign_9603b1() {
+  ivec3 res = ivec3(1);
+}
+
+void compute_main() {
+  sign_9603b1();
+}
+
+layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
+void main() {
+  compute_main();
+  return;
+}

test/tint/builtins/gen/literal/sign/9603b1.wgsl.expected.msl

@@ -0,0 +1,33 @@
+#include <metal_stdlib>
+
+using namespace metal;
+void sign_9603b1() {
+  int3 res = int3(1);
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  sign_9603b1();
+  return float4(0.0f);
+}
+
+vertex tint_symbol vertex_main() {
+  float4 const inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = {};
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+fragment void fragment_main() {
+  sign_9603b1();
+  return;
+}
+
+kernel void compute_main() {
+  sign_9603b1();
+  return;
+}
+

test/tint/builtins/gen/literal/sign/9603b1.wgsl.expected.spvasm

@@ -0,0 +1,68 @@
+; SPIR-V
+; Version: 1.3
+; Generator: Google Tint Compiler; 0
+; Bound: 34
+; Schema: 0
+               OpCapability Shader
+               OpMemoryModel Logical GLSL450
+               OpEntryPoint Vertex %vertex_main "vertex_main" %value %vertex_point_size
+               OpEntryPoint Fragment %fragment_main "fragment_main"
+               OpEntryPoint GLCompute %compute_main "compute_main"
+               OpExecutionMode %fragment_main OriginUpperLeft
+               OpExecutionMode %compute_main LocalSize 1 1 1
+               OpName %value "value"
+               OpName %vertex_point_size "vertex_point_size"
+               OpName %sign_9603b1 "sign_9603b1"
+               OpName %res "res"
+               OpName %vertex_main_inner "vertex_main_inner"
+               OpName %vertex_main "vertex_main"
+               OpName %fragment_main "fragment_main"
+               OpName %compute_main "compute_main"
+               OpDecorate %value BuiltIn Position
+               OpDecorate %vertex_point_size BuiltIn PointSize
+      %float = OpTypeFloat 32
+    %v4float = OpTypeVector %float 4
+%_ptr_Output_v4float = OpTypePointer Output %v4float
+          %5 = OpConstantNull %v4float
+      %value = OpVariable %_ptr_Output_v4float Output %5
+%_ptr_Output_float = OpTypePointer Output %float
+          %8 = OpConstantNull %float
+%vertex_point_size = OpVariable %_ptr_Output_float Output %8
+       %void = OpTypeVoid
+          %9 = OpTypeFunction %void
+        %int = OpTypeInt 32 1
+      %v3int = OpTypeVector %int 3
+      %int_1 = OpConstant %int 1
+         %16 = OpConstantComposite %v3int %int_1 %int_1 %int_1
+%_ptr_Function_v3int = OpTypePointer Function %v3int
+         %19 = OpConstantNull %v3int
+         %20 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%sign_9603b1 = OpFunction %void None %9
+         %12 = OpLabel
+        %res = OpVariable %_ptr_Function_v3int Function %19
+               OpStore %res %16
+               OpReturn
+               OpFunctionEnd
+%vertex_main_inner = OpFunction %v4float None %20
+         %22 = OpLabel
+         %23 = OpFunctionCall %void %sign_9603b1
+               OpReturnValue %5
+               OpFunctionEnd
+%vertex_main = OpFunction %void None %9
+         %25 = OpLabel
+         %26 = OpFunctionCall %v4float %vertex_main_inner
+               OpStore %value %26
+               OpStore %vertex_point_size %float_1
+               OpReturn
+               OpFunctionEnd
+%fragment_main = OpFunction %void None %9
+         %29 = OpLabel
+         %30 = OpFunctionCall %void %sign_9603b1
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %32 = OpLabel
+         %33 = OpFunctionCall %void %sign_9603b1
+               OpReturn
+               OpFunctionEnd

test/tint/builtins/gen/literal/sign/9603b1.wgsl.expected.wgsl

@@ -0,0 +1,19 @@
+fn sign_9603b1() {
+  var res : vec3<i32> = sign(vec3<i32>(1i));
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  sign_9603b1();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  sign_9603b1();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  sign_9603b1();
+}

test/tint/builtins/gen/var/sign/3233fa.wgsl

@@ -0,0 +1,44 @@
+// Copyright 2022 The Tint Authors.
+//
+// Licensed under the Apache License, Version 2.0 (the "License");
+// you may not use this file except in compliance with the License.
+// You may obtain a copy of the License at
+//
+//     http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing, software
+// distributed under the License is distributed on an "AS IS" BASIS,
+// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+// See the License for the specific language governing permissions and
+// limitations under the License.
+
+////////////////////////////////////////////////////////////////////////////////
+// File generated by tools/src/cmd/gen
+// using the template:
+//   test/tint/builtins/gen/gen.wgsl.tmpl
+//
+// Do not modify this file directly
+////////////////////////////////////////////////////////////////////////////////
+
+
+// fn sign(i32) -> i32
+fn sign_3233fa() {
+  var arg_0 = 1i;
+  var res: i32 = sign(arg_0);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  sign_3233fa();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  sign_3233fa();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  sign_3233fa();
+}

test/tint/builtins/gen/var/sign/3233fa.wgsl.expected.dxc.hlsl

@@ -0,0 +1,31 @@
+void sign_3233fa() {
+  int arg_0 = 1;
+  int res = sign(arg_0);
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  sign_3233fa();
+  return (0.0f).xxxx;
+}
+
+tint_symbol vertex_main() {
+  const float4 inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = (tint_symbol)0;
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+void fragment_main() {
+  sign_3233fa();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  sign_3233fa();
+  return;
+}

test/tint/builtins/gen/var/sign/3233fa.wgsl.expected.fxc.hlsl

@@ -0,0 +1,31 @@
+void sign_3233fa() {
+  int arg_0 = 1;
+  int res = sign(arg_0);
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  sign_3233fa();
+  return (0.0f).xxxx;
+}
+
+tint_symbol vertex_main() {
+  const float4 inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = (tint_symbol)0;
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+void fragment_main() {
+  sign_3233fa();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  sign_3233fa();
+  return;
+}

test/tint/builtins/gen/var/sign/3233fa.wgsl.expected.glsl

@@ -0,0 +1,52 @@
+#version 310 es
+
+void sign_3233fa() {
+  int arg_0 = 1;
+  int res = sign(arg_0);
+}
+
+vec4 vertex_main() {
+  sign_3233fa();
+  return vec4(0.0f);
+}
+
+void main() {
+  gl_PointSize = 1.0;
+  vec4 inner_result = vertex_main();
+  gl_Position = inner_result;
+  gl_Position.y = -(gl_Position.y);
+  gl_Position.z = ((2.0f * gl_Position.z) - gl_Position.w);
+  return;
+}
+#version 310 es
+precision mediump float;
+
+void sign_3233fa() {
+  int arg_0 = 1;
+  int res = sign(arg_0);
+}
+
+void fragment_main() {
+  sign_3233fa();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+void sign_3233fa() {
+  int arg_0 = 1;
+  int res = sign(arg_0);
+}
+
+void compute_main() {
+  sign_3233fa();
+}
+
+layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
+void main() {
+  compute_main();
+  return;
+}

test/tint/builtins/gen/var/sign/3233fa.wgsl.expected.msl

@@ -0,0 +1,38 @@
+#include <metal_stdlib>
+
+using namespace metal;
+int tint_sign(int v) {
+  return select(select(-1, 1, (v > 0)), 0, (v == 0));
+}
+
+void sign_3233fa() {
+  int arg_0 = 1;
+  int res = tint_sign(arg_0);
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  sign_3233fa();
+  return float4(0.0f);
+}
+
+vertex tint_symbol vertex_main() {
+  float4 const inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = {};
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+fragment void fragment_main() {
+  sign_3233fa();
+  return;
+}
+
+kernel void compute_main() {
+  sign_3233fa();
+  return;
+}
+

test/tint/builtins/gen/var/sign/3233fa.wgsl.expected.spvasm

@@ -0,0 +1,72 @@
+; SPIR-V
+; Version: 1.3
+; Generator: Google Tint Compiler; 0
+; Bound: 36
+; Schema: 0
+               OpCapability Shader
+         %19 = OpExtInstImport "GLSL.std.450"
+               OpMemoryModel Logical GLSL450
+               OpEntryPoint Vertex %vertex_main "vertex_main" %value %vertex_point_size
+               OpEntryPoint Fragment %fragment_main "fragment_main"
+               OpEntryPoint GLCompute %compute_main "compute_main"
+               OpExecutionMode %fragment_main OriginUpperLeft
+               OpExecutionMode %compute_main LocalSize 1 1 1
+               OpName %value "value"
+               OpName %vertex_point_size "vertex_point_size"
+               OpName %sign_3233fa "sign_3233fa"
+               OpName %arg_0 "arg_0"
+               OpName %res "res"
+               OpName %vertex_main_inner "vertex_main_inner"
+               OpName %vertex_main "vertex_main"
+               OpName %fragment_main "fragment_main"
+               OpName %compute_main "compute_main"
+               OpDecorate %value BuiltIn Position
+               OpDecorate %vertex_point_size BuiltIn PointSize
+      %float = OpTypeFloat 32
+    %v4float = OpTypeVector %float 4
+%_ptr_Output_v4float = OpTypePointer Output %v4float
+          %5 = OpConstantNull %v4float
+      %value = OpVariable %_ptr_Output_v4float Output %5
+%_ptr_Output_float = OpTypePointer Output %float
+          %8 = OpConstantNull %float
+%vertex_point_size = OpVariable %_ptr_Output_float Output %8
+       %void = OpTypeVoid
+          %9 = OpTypeFunction %void
+        %int = OpTypeInt 32 1
+      %int_1 = OpConstant %int 1
+%_ptr_Function_int = OpTypePointer Function %int
+         %17 = OpConstantNull %int
+         %22 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%sign_3233fa = OpFunction %void None %9
+         %12 = OpLabel
+      %arg_0 = OpVariable %_ptr_Function_int Function %17
+        %res = OpVariable %_ptr_Function_int Function %17
+               OpStore %arg_0 %int_1
+         %20 = OpLoad %int %arg_0
+         %18 = OpExtInst %int %19 SSign %20
+               OpStore %res %18
+               OpReturn
+               OpFunctionEnd
+%vertex_main_inner = OpFunction %v4float None %22
+         %24 = OpLabel
+         %25 = OpFunctionCall %void %sign_3233fa
+               OpReturnValue %5
+               OpFunctionEnd
+%vertex_main = OpFunction %void None %9
+         %27 = OpLabel
+         %28 = OpFunctionCall %v4float %vertex_main_inner
+               OpStore %value %28
+               OpStore %vertex_point_size %float_1
+               OpReturn
+               OpFunctionEnd
+%fragment_main = OpFunction %void None %9
+         %31 = OpLabel
+         %32 = OpFunctionCall %void %sign_3233fa
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %34 = OpLabel
+         %35 = OpFunctionCall %void %sign_3233fa
+               OpReturn
+               OpFunctionEnd

test/tint/builtins/gen/var/sign/3233fa.wgsl.expected.wgsl

@@ -0,0 +1,20 @@
+fn sign_3233fa() {
+  var arg_0 = 1i;
+  var res : i32 = sign(arg_0);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  sign_3233fa();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  sign_3233fa();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  sign_3233fa();
+}

test/tint/builtins/gen/var/sign/3a39ac.wgsl

@@ -0,0 +1,44 @@
+// Copyright 2022 The Tint Authors.
+//
+// Licensed under the Apache License, Version 2.0 (the "License");
+// you may not use this file except in compliance with the License.
+// You may obtain a copy of the License at
+//
+//     http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing, software
+// distributed under the License is distributed on an "AS IS" BASIS,
+// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+// See the License for the specific language governing permissions and
+// limitations under the License.
+
+////////////////////////////////////////////////////////////////////////////////
+// File generated by tools/src/cmd/gen
+// using the template:
+//   test/tint/builtins/gen/gen.wgsl.tmpl
+//
+// Do not modify this file directly
+////////////////////////////////////////////////////////////////////////////////
+
+
+// fn sign(ia) -> ia
+fn sign_3a39ac() {
+  const arg_0 = 1;
+  var res = sign(arg_0);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  sign_3a39ac();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  sign_3a39ac();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  sign_3a39ac();
+}

test/tint/builtins/gen/var/sign/3a39ac.wgsl.expected.dxc.hlsl

@@ -0,0 +1,30 @@
+void sign_3a39ac() {
+  int res = 1;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  sign_3a39ac();
+  return (0.0f).xxxx;
+}
+
+tint_symbol vertex_main() {
+  const float4 inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = (tint_symbol)0;
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+void fragment_main() {
+  sign_3a39ac();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  sign_3a39ac();
+  return;
+}

test/tint/builtins/gen/var/sign/3a39ac.wgsl.expected.fxc.hlsl

@@ -0,0 +1,30 @@
+void sign_3a39ac() {
+  int res = 1;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  sign_3a39ac();
+  return (0.0f).xxxx;
+}
+
+tint_symbol vertex_main() {
+  const float4 inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = (tint_symbol)0;
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+void fragment_main() {
+  sign_3a39ac();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  sign_3a39ac();
+  return;
+}

test/tint/builtins/gen/var/sign/3a39ac.wgsl.expected.glsl

@@ -0,0 +1,49 @@
+#version 310 es
+
+void sign_3a39ac() {
+  int res = 1;
+}
+
+vec4 vertex_main() {
+  sign_3a39ac();
+  return vec4(0.0f);
+}
+
+void main() {
+  gl_PointSize = 1.0;
+  vec4 inner_result = vertex_main();
+  gl_Position = inner_result;
+  gl_Position.y = -(gl_Position.y);
+  gl_Position.z = ((2.0f * gl_Position.z) - gl_Position.w);
+  return;
+}
+#version 310 es
+precision mediump float;
+
+void sign_3a39ac() {
+  int res = 1;
+}
+
+void fragment_main() {
+  sign_3a39ac();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+void sign_3a39ac() {
+  int res = 1;
+}
+
+void compute_main() {
+  sign_3a39ac();
+}
+
+layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
+void main() {
+  compute_main();
+  return;
+}

test/tint/builtins/gen/var/sign/3a39ac.wgsl.expected.msl

@@ -0,0 +1,33 @@
+#include <metal_stdlib>
+
+using namespace metal;
+void sign_3a39ac() {
+  int res = 1;
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  sign_3a39ac();
+  return float4(0.0f);
+}
+
+vertex tint_symbol vertex_main() {
+  float4 const inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = {};
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+fragment void fragment_main() {
+  sign_3a39ac();
+  return;
+}
+
+kernel void compute_main() {
+  sign_3a39ac();
+  return;
+}
+

test/tint/builtins/gen/var/sign/3a39ac.wgsl.expected.spvasm

@@ -0,0 +1,66 @@
+; SPIR-V
+; Version: 1.3
+; Generator: Google Tint Compiler; 0
+; Bound: 32
+; Schema: 0
+               OpCapability Shader
+               OpMemoryModel Logical GLSL450
+               OpEntryPoint Vertex %vertex_main "vertex_main" %value %vertex_point_size
+               OpEntryPoint Fragment %fragment_main "fragment_main"
+               OpEntryPoint GLCompute %compute_main "compute_main"
+               OpExecutionMode %fragment_main OriginUpperLeft
+               OpExecutionMode %compute_main LocalSize 1 1 1
+               OpName %value "value"
+               OpName %vertex_point_size "vertex_point_size"
+               OpName %sign_3a39ac "sign_3a39ac"
+               OpName %res "res"
+               OpName %vertex_main_inner "vertex_main_inner"
+               OpName %vertex_main "vertex_main"
+               OpName %fragment_main "fragment_main"
+               OpName %compute_main "compute_main"
+               OpDecorate %value BuiltIn Position
+               OpDecorate %vertex_point_size BuiltIn PointSize
+      %float = OpTypeFloat 32
+    %v4float = OpTypeVector %float 4
+%_ptr_Output_v4float = OpTypePointer Output %v4float
+          %5 = OpConstantNull %v4float
+      %value = OpVariable %_ptr_Output_v4float Output %5
+%_ptr_Output_float = OpTypePointer Output %float
+          %8 = OpConstantNull %float
+%vertex_point_size = OpVariable %_ptr_Output_float Output %8
+       %void = OpTypeVoid
+          %9 = OpTypeFunction %void
+        %int = OpTypeInt 32 1
+      %int_1 = OpConstant %int 1
+%_ptr_Function_int = OpTypePointer Function %int
+         %17 = OpConstantNull %int
+         %18 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%sign_3a39ac = OpFunction %void None %9
+         %12 = OpLabel
+        %res = OpVariable %_ptr_Function_int Function %17
+               OpStore %res %int_1
+               OpReturn
+               OpFunctionEnd
+%vertex_main_inner = OpFunction %v4float None %18
+         %20 = OpLabel
+         %21 = OpFunctionCall %void %sign_3a39ac
+               OpReturnValue %5
+               OpFunctionEnd
+%vertex_main = OpFunction %void None %9
+         %23 = OpLabel
+         %24 = OpFunctionCall %v4float %vertex_main_inner
+               OpStore %value %24
+               OpStore %vertex_point_size %float_1
+               OpReturn
+               OpFunctionEnd
+%fragment_main = OpFunction %void None %9
+         %27 = OpLabel
+         %28 = OpFunctionCall %void %sign_3a39ac
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %30 = OpLabel
+         %31 = OpFunctionCall %void %sign_3a39ac
+               OpReturn
+               OpFunctionEnd

test/tint/builtins/gen/var/sign/3a39ac.wgsl.expected.wgsl

@@ -0,0 +1,20 @@
+fn sign_3a39ac() {
+  const arg_0 = 1;
+  var res = sign(arg_0);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  sign_3a39ac();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  sign_3a39ac();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  sign_3a39ac();
+}

test/tint/builtins/gen/var/sign/3bdab6.wgsl

@@ -0,0 +1,44 @@
+// Copyright 2022 The Tint Authors.
+//
+// Licensed under the Apache License, Version 2.0 (the "License");
+// you may not use this file except in compliance with the License.
+// You may obtain a copy of the License at
+//
+//     http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing, software
+// distributed under the License is distributed on an "AS IS" BASIS,
+// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+// See the License for the specific language governing permissions and
+// limitations under the License.
+
+////////////////////////////////////////////////////////////////////////////////
+// File generated by tools/src/cmd/gen
+// using the template:
+//   test/tint/builtins/gen/gen.wgsl.tmpl
+//
+// Do not modify this file directly
+////////////////////////////////////////////////////////////////////////////////
+
+
+// fn sign(vec<4, ia>) -> vec<4, ia>
+fn sign_3bdab6() {
+  const arg_0 = vec4(1);
+  var res = sign(arg_0);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  sign_3bdab6();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  sign_3bdab6();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  sign_3bdab6();
+}

test/tint/builtins/gen/var/sign/3bdab6.wgsl.expected.dxc.hlsl

@@ -0,0 +1,30 @@
+void sign_3bdab6() {
+  int4 res = (1).xxxx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  sign_3bdab6();
+  return (0.0f).xxxx;
+}
+
+tint_symbol vertex_main() {
+  const float4 inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = (tint_symbol)0;
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+void fragment_main() {
+  sign_3bdab6();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  sign_3bdab6();
+  return;
+}

test/tint/builtins/gen/var/sign/3bdab6.wgsl.expected.fxc.hlsl

@@ -0,0 +1,30 @@
+void sign_3bdab6() {
+  int4 res = (1).xxxx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  sign_3bdab6();
+  return (0.0f).xxxx;
+}
+
+tint_symbol vertex_main() {
+  const float4 inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = (tint_symbol)0;
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+void fragment_main() {
+  sign_3bdab6();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  sign_3bdab6();
+  return;
+}

test/tint/builtins/gen/var/sign/3bdab6.wgsl.expected.glsl

@@ -0,0 +1,49 @@
+#version 310 es
+
+void sign_3bdab6() {
+  ivec4 res = ivec4(1);
+}
+
+vec4 vertex_main() {
+  sign_3bdab6();
+  return vec4(0.0f);
+}
+
+void main() {
+  gl_PointSize = 1.0;
+  vec4 inner_result = vertex_main();
+  gl_Position = inner_result;
+  gl_Position.y = -(gl_Position.y);
+  gl_Position.z = ((2.0f * gl_Position.z) - gl_Position.w);
+  return;
+}
+#version 310 es
+precision mediump float;
+
+void sign_3bdab6() {
+  ivec4 res = ivec4(1);
+}
+
+void fragment_main() {
+  sign_3bdab6();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+void sign_3bdab6() {
+  ivec4 res = ivec4(1);
+}
+
+void compute_main() {
+  sign_3bdab6();
+}
+
+layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
+void main() {
+  compute_main();
+  return;
+}

test/tint/builtins/gen/var/sign/3bdab6.wgsl.expected.msl

@@ -0,0 +1,33 @@
+#include <metal_stdlib>
+
+using namespace metal;
+void sign_3bdab6() {
+  int4 res = int4(1);
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  sign_3bdab6();
+  return float4(0.0f);
+}
+
+vertex tint_symbol vertex_main() {
+  float4 const inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = {};
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+fragment void fragment_main() {
+  sign_3bdab6();
+  return;
+}
+
+kernel void compute_main() {
+  sign_3bdab6();
+  return;
+}
+

test/tint/builtins/gen/var/sign/3bdab6.wgsl.expected.spvasm

@@ -0,0 +1,68 @@
+; SPIR-V
+; Version: 1.3
+; Generator: Google Tint Compiler; 0
+; Bound: 34
+; Schema: 0
+               OpCapability Shader
+               OpMemoryModel Logical GLSL450
+               OpEntryPoint Vertex %vertex_main "vertex_main" %value %vertex_point_size
+               OpEntryPoint Fragment %fragment_main "fragment_main"
+               OpEntryPoint GLCompute %compute_main "compute_main"
+               OpExecutionMode %fragment_main OriginUpperLeft
+               OpExecutionMode %compute_main LocalSize 1 1 1
+               OpName %value "value"
+               OpName %vertex_point_size "vertex_point_size"
+               OpName %sign_3bdab6 "sign_3bdab6"
+               OpName %res "res"
+               OpName %vertex_main_inner "vertex_main_inner"
+               OpName %vertex_main "vertex_main"
+               OpName %fragment_main "fragment_main"
+               OpName %compute_main "compute_main"
+               OpDecorate %value BuiltIn Position
+               OpDecorate %vertex_point_size BuiltIn PointSize
+      %float = OpTypeFloat 32
+    %v4float = OpTypeVector %float 4
+%_ptr_Output_v4float = OpTypePointer Output %v4float
+          %5 = OpConstantNull %v4float
+      %value = OpVariable %_ptr_Output_v4float Output %5
+%_ptr_Output_float = OpTypePointer Output %float
+          %8 = OpConstantNull %float
+%vertex_point_size = OpVariable %_ptr_Output_float Output %8
+       %void = OpTypeVoid
+          %9 = OpTypeFunction %void
+        %int = OpTypeInt 32 1
+      %v4int = OpTypeVector %int 4
+      %int_1 = OpConstant %int 1
+         %16 = OpConstantComposite %v4int %int_1 %int_1 %int_1 %int_1
+%_ptr_Function_v4int = OpTypePointer Function %v4int
+         %19 = OpConstantNull %v4int
+         %20 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%sign_3bdab6 = OpFunction %void None %9
+         %12 = OpLabel
+        %res = OpVariable %_ptr_Function_v4int Function %19
+               OpStore %res %16
+               OpReturn
+               OpFunctionEnd
+%vertex_main_inner = OpFunction %v4float None %20
+         %22 = OpLabel
+         %23 = OpFunctionCall %void %sign_3bdab6
+               OpReturnValue %5
+               OpFunctionEnd
+%vertex_main = OpFunction %void None %9
+         %25 = OpLabel
+         %26 = OpFunctionCall %v4float %vertex_main_inner
+               OpStore %value %26
+               OpStore %vertex_point_size %float_1
+               OpReturn
+               OpFunctionEnd
+%fragment_main = OpFunction %void None %9
+         %29 = OpLabel
+         %30 = OpFunctionCall %void %sign_3bdab6
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %32 = OpLabel
+         %33 = OpFunctionCall %void %sign_3bdab6
+               OpReturn
+               OpFunctionEnd

test/tint/builtins/gen/var/sign/3bdab6.wgsl.expected.wgsl

@@ -0,0 +1,20 @@
+fn sign_3bdab6() {
+  const arg_0 = vec4(1);
+  var res = sign(arg_0);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  sign_3bdab6();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  sign_3bdab6();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  sign_3bdab6();
+}

test/tint/builtins/gen/var/sign/55339e.wgsl

@@ -0,0 +1,44 @@
+// Copyright 2022 The Tint Authors.
+//
+// Licensed under the Apache License, Version 2.0 (the "License");
+// you may not use this file except in compliance with the License.
+// You may obtain a copy of the License at
+//
+//     http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing, software
+// distributed under the License is distributed on an "AS IS" BASIS,
+// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+// See the License for the specific language governing permissions and
+// limitations under the License.
+
+////////////////////////////////////////////////////////////////////////////////
+// File generated by tools/src/cmd/gen
+// using the template:
+//   test/tint/builtins/gen/gen.wgsl.tmpl
+//
+// Do not modify this file directly
+////////////////////////////////////////////////////////////////////////////////
+
+
+// fn sign(vec<3, ia>) -> vec<3, ia>
+fn sign_55339e() {
+  const arg_0 = vec3(1);
+  var res = sign(arg_0);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  sign_55339e();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  sign_55339e();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  sign_55339e();
+}

test/tint/builtins/gen/var/sign/55339e.wgsl.expected.dxc.hlsl

@@ -0,0 +1,30 @@
+void sign_55339e() {
+  int3 res = (1).xxx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  sign_55339e();
+  return (0.0f).xxxx;
+}
+
+tint_symbol vertex_main() {
+  const float4 inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = (tint_symbol)0;
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+void fragment_main() {
+  sign_55339e();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  sign_55339e();
+  return;
+}

test/tint/builtins/gen/var/sign/55339e.wgsl.expected.fxc.hlsl

@@ -0,0 +1,30 @@
+void sign_55339e() {
+  int3 res = (1).xxx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  sign_55339e();
+  return (0.0f).xxxx;
+}
+
+tint_symbol vertex_main() {
+  const float4 inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = (tint_symbol)0;
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+void fragment_main() {
+  sign_55339e();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  sign_55339e();
+  return;
+}

test/tint/builtins/gen/var/sign/55339e.wgsl.expected.glsl

@@ -0,0 +1,49 @@
+#version 310 es
+
+void sign_55339e() {
+  ivec3 res = ivec3(1);
+}
+
+vec4 vertex_main() {
+  sign_55339e();
+  return vec4(0.0f);
+}
+
+void main() {
+  gl_PointSize = 1.0;
+  vec4 inner_result = vertex_main();
+  gl_Position = inner_result;
+  gl_Position.y = -(gl_Position.y);
+  gl_Position.z = ((2.0f * gl_Position.z) - gl_Position.w);
+  return;
+}
+#version 310 es
+precision mediump float;
+
+void sign_55339e() {
+  ivec3 res = ivec3(1);
+}
+
+void fragment_main() {
+  sign_55339e();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+void sign_55339e() {
+  ivec3 res = ivec3(1);
+}
+
+void compute_main() {
+  sign_55339e();
+}
+
+layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
+void main() {
+  compute_main();
+  return;
+}

test/tint/builtins/gen/var/sign/55339e.wgsl.expected.msl

@@ -0,0 +1,33 @@
+#include <metal_stdlib>
+
+using namespace metal;
+void sign_55339e() {
+  int3 res = int3(1);
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  sign_55339e();
+  return float4(0.0f);
+}
+
+vertex tint_symbol vertex_main() {
+  float4 const inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = {};
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+fragment void fragment_main() {
+  sign_55339e();
+  return;
+}
+
+kernel void compute_main() {
+  sign_55339e();
+  return;
+}
+

test/tint/builtins/gen/var/sign/55339e.wgsl.expected.spvasm

@@ -0,0 +1,68 @@
+; SPIR-V
+; Version: 1.3
+; Generator: Google Tint Compiler; 0
+; Bound: 34
+; Schema: 0
+               OpCapability Shader
+               OpMemoryModel Logical GLSL450
+               OpEntryPoint Vertex %vertex_main "vertex_main" %value %vertex_point_size
+               OpEntryPoint Fragment %fragment_main "fragment_main"
+               OpEntryPoint GLCompute %compute_main "compute_main"
+               OpExecutionMode %fragment_main OriginUpperLeft
+               OpExecutionMode %compute_main LocalSize 1 1 1
+               OpName %value "value"
+               OpName %vertex_point_size "vertex_point_size"
+               OpName %sign_55339e "sign_55339e"
+               OpName %res "res"
+               OpName %vertex_main_inner "vertex_main_inner"
+               OpName %vertex_main "vertex_main"
+               OpName %fragment_main "fragment_main"
+               OpName %compute_main "compute_main"
+               OpDecorate %value BuiltIn Position
+               OpDecorate %vertex_point_size BuiltIn PointSize
+      %float = OpTypeFloat 32
+    %v4float = OpTypeVector %float 4
+%_ptr_Output_v4float = OpTypePointer Output %v4float
+          %5 = OpConstantNull %v4float
+      %value = OpVariable %_ptr_Output_v4float Output %5
+%_ptr_Output_float = OpTypePointer Output %float
+          %8 = OpConstantNull %float
+%vertex_point_size = OpVariable %_ptr_Output_float Output %8
+       %void = OpTypeVoid
+          %9 = OpTypeFunction %void
+        %int = OpTypeInt 32 1
+      %v3int = OpTypeVector %int 3
+      %int_1 = OpConstant %int 1
+         %16 = OpConstantComposite %v3int %int_1 %int_1 %int_1
+%_ptr_Function_v3int = OpTypePointer Function %v3int
+         %19 = OpConstantNull %v3int
+         %20 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%sign_55339e = OpFunction %void None %9
+         %12 = OpLabel
+        %res = OpVariable %_ptr_Function_v3int Function %19
+               OpStore %res %16
+               OpReturn
+               OpFunctionEnd
+%vertex_main_inner = OpFunction %v4float None %20
+         %22 = OpLabel
+         %23 = OpFunctionCall %void %sign_55339e
+               OpReturnValue %5
+               OpFunctionEnd
+%vertex_main = OpFunction %void None %9
+         %25 = OpLabel
+         %26 = OpFunctionCall %v4float %vertex_main_inner
+               OpStore %value %26
+               OpStore %vertex_point_size %float_1
+               OpReturn
+               OpFunctionEnd
+%fragment_main = OpFunction %void None %9
+         %29 = OpLabel
+         %30 = OpFunctionCall %void %sign_55339e
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %32 = OpLabel
+         %33 = OpFunctionCall %void %sign_55339e
+               OpReturn
+               OpFunctionEnd

test/tint/builtins/gen/var/sign/55339e.wgsl.expected.wgsl

@@ -0,0 +1,20 @@
+fn sign_55339e() {
+  const arg_0 = vec3(1);
+  var res = sign(arg_0);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  sign_55339e();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  sign_55339e();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  sign_55339e();
+}

test/tint/builtins/gen/var/sign/58d779.wgsl

@@ -0,0 +1,44 @@
+// Copyright 2022 The Tint Authors.
+//
+// Licensed under the Apache License, Version 2.0 (the "License");
+// you may not use this file except in compliance with the License.
+// You may obtain a copy of the License at
+//
+//     http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing, software
+// distributed under the License is distributed on an "AS IS" BASIS,
+// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+// See the License for the specific language governing permissions and
+// limitations under the License.
+
+////////////////////////////////////////////////////////////////////////////////
+// File generated by tools/src/cmd/gen
+// using the template:
+//   test/tint/builtins/gen/gen.wgsl.tmpl
+//
+// Do not modify this file directly
+////////////////////////////////////////////////////////////////////////////////
+
+
+// fn sign(vec<4, i32>) -> vec<4, i32>
+fn sign_58d779() {
+  var arg_0 = vec4<i32>(1i);
+  var res: vec4<i32> = sign(arg_0);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  sign_58d779();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  sign_58d779();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  sign_58d779();
+}

test/tint/builtins/gen/var/sign/58d779.wgsl.expected.dxc.hlsl

@@ -0,0 +1,31 @@
+void sign_58d779() {
+  int4 arg_0 = (1).xxxx;
+  int4 res = sign(arg_0);
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  sign_58d779();
+  return (0.0f).xxxx;
+}
+
+tint_symbol vertex_main() {
+  const float4 inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = (tint_symbol)0;
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+void fragment_main() {
+  sign_58d779();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  sign_58d779();
+  return;
+}

test/tint/builtins/gen/var/sign/58d779.wgsl.expected.fxc.hlsl

@@ -0,0 +1,31 @@
+void sign_58d779() {
+  int4 arg_0 = (1).xxxx;
+  int4 res = sign(arg_0);
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  sign_58d779();
+  return (0.0f).xxxx;
+}
+
+tint_symbol vertex_main() {
+  const float4 inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = (tint_symbol)0;
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+void fragment_main() {
+  sign_58d779();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  sign_58d779();
+  return;
+}

test/tint/builtins/gen/var/sign/58d779.wgsl.expected.glsl

@@ -0,0 +1,52 @@
+#version 310 es
+
+void sign_58d779() {
+  ivec4 arg_0 = ivec4(1);
+  ivec4 res = sign(arg_0);
+}
+
+vec4 vertex_main() {
+  sign_58d779();
+  return vec4(0.0f);
+}
+
+void main() {
+  gl_PointSize = 1.0;
+  vec4 inner_result = vertex_main();
+  gl_Position = inner_result;
+  gl_Position.y = -(gl_Position.y);
+  gl_Position.z = ((2.0f * gl_Position.z) - gl_Position.w);
+  return;
+}
+#version 310 es
+precision mediump float;
+
+void sign_58d779() {
+  ivec4 arg_0 = ivec4(1);
+  ivec4 res = sign(arg_0);
+}
+
+void fragment_main() {
+  sign_58d779();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+void sign_58d779() {
+  ivec4 arg_0 = ivec4(1);
+  ivec4 res = sign(arg_0);
+}
+
+void compute_main() {
+  sign_58d779();
+}
+
+layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
+void main() {
+  compute_main();
+  return;
+}

test/tint/builtins/gen/var/sign/58d779.wgsl.expected.msl

@@ -0,0 +1,38 @@
+#include <metal_stdlib>
+
+using namespace metal;
+int4 tint_sign(int4 v) {
+  return select(select(int4(-1), int4(1), (v > int4(0))), int4(0), (v == int4(0)));
+}
+
+void sign_58d779() {
+  int4 arg_0 = int4(1);
+  int4 res = tint_sign(arg_0);
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  sign_58d779();
+  return float4(0.0f);
+}
+
+vertex tint_symbol vertex_main() {
+  float4 const inner_result = vertex_main_inner();
+  tint_symbol wrapper_result = {};
+  wrapper_result.value = inner_result;
+  return wrapper_result;
+}
+
+fragment void fragment_main() {
+  sign_58d779();
+  return;
+}
+
+kernel void compute_main() {
+  sign_58d779();
+  return;
+}
+

test/tint/builtins/gen/var/sign/58d779.wgsl.expected.spvasm

@@ -0,0 +1,74 @@
+; SPIR-V
+; Version: 1.3
+; Generator: Google Tint Compiler; 0
+; Bound: 38
+; Schema: 0
+               OpCapability Shader
+         %21 = OpExtInstImport "GLSL.std.450"
+               OpMemoryModel Logical GLSL450
+               OpEntryPoint Vertex %vertex_main "vertex_main" %value %vertex_point_size
+               OpEntryPoint Fragment %fragment_main "fragment_main"
+               OpEntryPoint GLCompute %compute_main "compute_main"
+               OpExecutionMode %fragment_main OriginUpperLeft
+               OpExecutionMode %compute_main LocalSize 1 1 1
+               OpName %value "value"
+               OpName %vertex_point_size "vertex_point_size"
+               OpName %sign_58d779 "sign_58d779"
+               OpName %arg_0 "arg_0"
+               OpName %res "res"
+               OpName %vertex_main_inner "vertex_main_inner"
+               OpName %vertex_main "vertex_main"
+               OpName %fragment_main "fragment_main"
+               OpName %compute_main "compute_main"
+               OpDecorate %value BuiltIn Position
+               OpDecorate %vertex_point_size BuiltIn PointSize
+      %float = OpTypeFloat 32
+    %v4float = OpTypeVector %float 4
+%_ptr_Output_v4float = OpTypePointer Output %v4float
+          %5 = OpConstantNull %v4float
+      %value = OpVariable %_ptr_Output_v4float Output %5
+%_ptr_Output_float = OpTypePointer Output %float
+          %8 = OpConstantNull %float
+%vertex_point_size = OpVariable %_ptr_Output_float Output %8
+       %void = OpTypeVoid
+          %9 = OpTypeFunction %void
+        %int = OpTypeInt 32 1
+      %v4int = OpTypeVector %int 4
+      %int_1 = OpConstant %int 1
+         %16 = OpConstantComposite %v4int %int_1 %int_1 %int_1 %int_1
+%_ptr_Function_v4int = OpTypePointer Function %v4int
+         %19 = OpConstantNull %v4int
+         %24 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%sign_58d779 = OpFunction %void None %9
+         %12 = OpLabel
+      %arg_0 = OpVariable %_ptr_Function_v4int Function %19
+        %res = OpVariable %_ptr_Function_v4int Function %19
+               OpStore %arg_0 %16
+         %22 = OpLoad %v4int %arg_0
+         %20 = OpExtInst %v4int %21 SSign %22
+               OpStore %res %20
+               OpReturn
+               OpFunctionEnd
+%vertex_main_inner = OpFunction %v4float None %24
+         %26 = OpLabel
+         %27 = OpFunctionCall %void %sign_58d779
+               OpReturnValue %5
+               OpFunctionEnd
+%vertex_main = OpFunction %void None %9
+         %29 = OpLabel
+         %30 = OpFunctionCall %v4float %vertex_main_inner
+               OpStore %value %30
+               OpStore %vertex_point_size %float_1
+               OpReturn
+               OpFunctionEnd
+%fragment_main = OpFunction %void None %9
+         %33 = OpLabel
+         %34 = OpFunctionCall %void %sign_58d779
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %36 = OpLabel
+         %37 = OpFunctionCall %void %sign_58d779
+               OpReturn
+               OpFunctionEnd

test/tint/builtins/gen/var/sign/58d779.wgsl.expected.wgsl

@@ -0,0 +1,20 @@
+fn sign_58d779() {
+  var arg_0 = vec4<i32>(1i);
+  var res : vec4<i32> = sign(arg_0);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  sign_58d779();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  sign_58d779();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  sign_58d779();
+}