diff --git a/src/tint/intrinsics.def b/src/tint/intrinsics.def
index a148d4d9c4..687b8e1026 100644
--- a/src/tint/intrinsics.def
+++ b/src/tint/intrinsics.def
@@ -466,7 +466,7 @@ fn dot4U8Packed(u32, u32) -> u32
 @const fn exp2<N: num, T: fa_f32_f16>(vec<N, T>) -> vec<N, T>
 @const fn extractBits<T: iu32>(T, u32, u32) -> T
 @const fn extractBits<N: num, T: iu32>(vec<N, T>, u32, u32) -> vec<N, T>
-fn faceForward<N: num, T: f32_f16>(vec<N, T>, vec<N, T>, vec<N, T>) -> vec<N, T>
+@const fn faceForward<N: num, T: fa_f32_f16>(vec<N, T>, vec<N, T>, vec<N, T>) -> vec<N, T>
 @const fn firstLeadingBit<T: iu32>(T) -> T
 @const fn firstLeadingBit<N: num, T: iu32>(vec<N, T>) -> vec<N, T>
 @const fn firstTrailingBit<T: iu32>(T) -> T
diff --git a/src/tint/resolver/const_eval.cc b/src/tint/resolver/const_eval.cc
index 159aaf4b0d..f1197ce8bc 100644
--- a/src/tint/resolver/const_eval.cc
+++ b/src/tint/resolver/const_eval.cc
@@ -2367,6 +2367,25 @@ ConstEval::Result ConstEval::extractBits(const sem::Type* ty,
     return TransformElements(builder, ty, transform, args[0]);
 }
 
+ConstEval::Result ConstEval::faceForward(const sem::Type* ty,
+                                         utils::VectorRef<const sem::Constant*> args,
+                                         const Source& source) {
+    // Returns e1 if dot(e2, e3) is negative, and -e1 otherwise.
+    auto* e1 = args[0];
+    auto* e2 = args[1];
+    auto* e3 = args[2];
+    auto r = Dot(source, e2, e3);
+    if (!r) {
+        AddNote("when calculating faceForward", source);
+        return utils::Failure;
+    }
+    auto is_negative = [](auto v) { return v < 0; };
+    if (Dispatch_fa_f32_f16(is_negative, r.Get())) {
+        return e1;
+    }
+    return OpUnaryMinus(ty, utils::Vector{e1}, source);
+}
+
 ConstEval::Result ConstEval::firstLeadingBit(const sem::Type* ty,
                                              utils::VectorRef<const sem::Constant*> args,
                                              const Source& source) {
diff --git a/src/tint/resolver/const_eval.h b/src/tint/resolver/const_eval.h
index 0c8ec6b758..df1db17806 100644
--- a/src/tint/resolver/const_eval.h
+++ b/src/tint/resolver/const_eval.h
@@ -602,6 +602,15 @@ class ConstEval {
                        utils::VectorRef<const sem::Constant*> args,
                        const Source& source);
 
+    /// faceForward builtin
+    /// @param ty the expression type
+    /// @param args the input arguments
+    /// @param source the source location
+    /// @return the result value, or null if the value cannot be calculated
+    Result faceForward(const sem::Type* ty,
+                       utils::VectorRef<const sem::Constant*> args,
+                       const Source& source);
+
     /// firstLeadingBit builtin
     /// @param ty the expression type
     /// @param args the input arguments
diff --git a/src/tint/resolver/const_eval_builtin_test.cc b/src/tint/resolver/const_eval_builtin_test.cc
index dabe7d954c..7c5d708ba2 100644
--- a/src/tint/resolver/const_eval_builtin_test.cc
+++ b/src/tint/resolver/const_eval_builtin_test.cc
@@ -953,6 +953,88 @@ INSTANTIATE_TEST_SUITE_P(  //
                                               DeterminantCases<f32>(),     //
                                               DeterminantCases<f16>()))));
 
+template <typename T>
+std::vector<Case> FaceForwardCases() {
+    // Rotate v by degs around Z axis
+    auto rotate = [&](const Value& v, float degs) {
+        auto x = builder::As<T>(v.args[0]);
+        auto y = builder::As<T>(v.args[1]);
+        auto z = builder::As<T>(v.args[2]);
+        auto rads = T(degs) * kPi<T> / T(180);
+        auto x2 = T(x * std::cos(rads) - y * std::sin(rads));
+        auto y2 = T(x * std::sin(rads) + y * std::cos(rads));
+        return Vec(x2, y2, z);
+    };
+
+    // An arbitrary input vector and its negation, used for e1 args to FaceForward
+    auto pos_vec = Vec(T(1), T(2), T(3));
+    auto neg_vec = Vec(-T(1), -T(2), -T(3));
+
+    // An arbitrary vector in the xy plane, used for e2 and e3 args to FaceForward.
+    auto fwd_xy = Vec(T(1.23), T(4.56), T(0));
+
+    std::vector<Case> r = {
+        C({pos_vec, fwd_xy, rotate(fwd_xy, 85)}, neg_vec),
+        C({pos_vec, fwd_xy, rotate(fwd_xy, 85)}, neg_vec),
+        C({pos_vec, fwd_xy, rotate(fwd_xy, 95)}, pos_vec),
+        C({pos_vec, fwd_xy, rotate(fwd_xy, -95)}, pos_vec),
+        C({pos_vec, fwd_xy, rotate(fwd_xy, 180)}, pos_vec),
+
+        C({pos_vec, rotate(fwd_xy, 33), rotate(fwd_xy, 33 + 85)}, neg_vec),
+        C({pos_vec, rotate(fwd_xy, 33), rotate(fwd_xy, 33 - 85)}, neg_vec),
+        C({pos_vec, rotate(fwd_xy, 33), rotate(fwd_xy, 33 + 95)}, pos_vec),
+        C({pos_vec, rotate(fwd_xy, 33), rotate(fwd_xy, 33 - 95)}, pos_vec),
+        C({pos_vec, rotate(fwd_xy, 33), rotate(fwd_xy, 33 + 180)}, pos_vec),
+
+        C({pos_vec, rotate(fwd_xy, 234), rotate(fwd_xy, 234 + 85)}, neg_vec),
+        C({pos_vec, rotate(fwd_xy, 234), rotate(fwd_xy, 234 - 85)}, neg_vec),
+        C({pos_vec, rotate(fwd_xy, 234), rotate(fwd_xy, 234 + 95)}, pos_vec),
+        C({pos_vec, rotate(fwd_xy, 234), rotate(fwd_xy, 234 - 95)}, pos_vec),
+        C({pos_vec, rotate(fwd_xy, 234), rotate(fwd_xy, 234 + 180)}, pos_vec),
+
+        // Same, but swap input and result vectors
+        C({neg_vec, fwd_xy, rotate(fwd_xy, 85)}, pos_vec),
+        C({neg_vec, fwd_xy, rotate(fwd_xy, 85)}, pos_vec),
+        C({neg_vec, fwd_xy, rotate(fwd_xy, 95)}, neg_vec),
+        C({neg_vec, fwd_xy, rotate(fwd_xy, -95)}, neg_vec),
+        C({neg_vec, fwd_xy, rotate(fwd_xy, 180)}, neg_vec),
+
+        C({neg_vec, rotate(fwd_xy, 33), rotate(fwd_xy, 33 + 85)}, pos_vec),
+        C({neg_vec, rotate(fwd_xy, 33), rotate(fwd_xy, 33 - 85)}, pos_vec),
+        C({neg_vec, rotate(fwd_xy, 33), rotate(fwd_xy, 33 + 95)}, neg_vec),
+        C({neg_vec, rotate(fwd_xy, 33), rotate(fwd_xy, 33 - 95)}, neg_vec),
+        C({neg_vec, rotate(fwd_xy, 33), rotate(fwd_xy, 33 + 180)}, neg_vec),
+
+        C({neg_vec, rotate(fwd_xy, 234), rotate(fwd_xy, 234 + 85)}, pos_vec),
+        C({neg_vec, rotate(fwd_xy, 234), rotate(fwd_xy, 234 - 85)}, pos_vec),
+        C({neg_vec, rotate(fwd_xy, 234), rotate(fwd_xy, 234 + 95)}, neg_vec),
+        C({neg_vec, rotate(fwd_xy, 234), rotate(fwd_xy, 234 - 95)}, neg_vec),
+        C({neg_vec, rotate(fwd_xy, 234), rotate(fwd_xy, 234 + 180)}, neg_vec),
+    };
+
+    auto error_msg = [](auto a, const char* op, auto b) {
+        return "12:34 error: " + OverflowErrorMessage(a, op, b) + R"(
+12:34 note: when calculating faceForward)";
+    };
+    ConcatInto(  //
+        r, std::vector<Case>{
+               // Overflow the dot product operation
+               E({pos_vec, Vec(T::Highest(), T::Highest(), T(0)), Vec(T(1), T(1), T(0))},
+                 error_msg(T::Highest(), "+", T::Highest())),
+               E({pos_vec, Vec(T::Lowest(), T::Lowest(), T(0)), Vec(T(1), T(1), T(0))},
+                 error_msg(T::Lowest(), "+", T::Lowest())),
+           });
+
+    return r;
+}
+INSTANTIATE_TEST_SUITE_P(  //
+    FaceForward,
+    ResolverConstEvalBuiltinTest,
+    testing::Combine(testing::Values(sem::BuiltinType::kFaceForward),
+                     testing::ValuesIn(Concat(FaceForwardCases<AFloat>(),  //
+                                              FaceForwardCases<f32>(),     //
+                                              FaceForwardCases<f16>()))));
+
 template <typename T>
 std::vector<Case> FirstLeadingBitCases() {
     using B = BitValues<T>;
diff --git a/src/tint/resolver/intrinsic_table.inl b/src/tint/resolver/intrinsic_table.inl
index 9d77df3a2f..e521a9ce00 100644
--- a/src/tint/resolver/intrinsic_table.inl
+++ b/src/tint/resolver/intrinsic_table.inl
@@ -13614,12 +13614,12 @@ constexpr OverloadInfo kOverloads[] = {
     /* num parameters */ 3,
     /* num template types */ 1,
     /* num template numbers */ 1,
-    /* template types */ &kTemplateTypes[26],
+    /* template types */ &kTemplateTypes[23],
     /* template numbers */ &kTemplateNumbers[4],
     /* parameters */ &kParameters[459],
     /* return matcher indices */ &kMatcherIndices[30],
     /* flags */ OverloadFlags(OverloadFlag::kIsBuiltin, OverloadFlag::kSupportsVertexPipeline, OverloadFlag::kSupportsFragmentPipeline, OverloadFlag::kSupportsComputePipeline),
-    /* const eval */ nullptr,
+    /* const eval */ &ConstEval::faceForward,
   },
   {
     /* [442] */
@@ -14218,7 +14218,7 @@ constexpr IntrinsicInfo kBuiltins[] = {
   },
   {
     /* [34] */
-    /* fn faceForward<N : num, T : f32_f16>(vec<N, T>, vec<N, T>, vec<N, T>) -> vec<N, T> */
+    /* fn faceForward<N : num, T : fa_f32_f16>(vec<N, T>, vec<N, T>, vec<N, T>) -> vec<N, T> */
     /* num overloads */ 1,
     /* overloads */ &kOverloads[441],
   },
diff --git a/src/tint/resolver/resolver_test_helper.h b/src/tint/resolver/resolver_test_helper.h
index cf17b616df..d44c9d81df 100644
--- a/src/tint/resolver/resolver_test_helper.h
+++ b/src/tint/resolver/resolver_test_helper.h
@@ -185,7 +185,7 @@ using Scalar = std::variant<i32, u32, f32, f16, AInt, AFloat, bool>;
 
 /// Returns current variant value in `s` cast to type `T`
 template <typename T>
-T As(Scalar& s) {
+T As(const Scalar& s) {
     return std::visit([](auto&& v) { return static_cast<T>(v); }, s);
 }
 
diff --git a/test/tint/builtins/gen/literal/faceForward/2c4d14.wgsl b/test/tint/builtins/gen/literal/faceForward/2c4d14.wgsl
new file mode 100644
index 0000000000..52173e74fa
--- /dev/null
+++ b/test/tint/builtins/gen/literal/faceForward/2c4d14.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 faceForward(vec<4, fa>, vec<4, fa>, vec<4, fa>) -> vec<4, fa>
+fn faceForward_2c4d14() {
+  var res = faceForward(vec4(1.), vec4(1.), vec4(1.));
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  faceForward_2c4d14();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  faceForward_2c4d14();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  faceForward_2c4d14();
+}
diff --git a/test/tint/builtins/gen/literal/faceForward/2c4d14.wgsl.expected.dxc.hlsl b/test/tint/builtins/gen/literal/faceForward/2c4d14.wgsl.expected.dxc.hlsl
new file mode 100644
index 0000000000..1109cebb9a
--- /dev/null
+++ b/test/tint/builtins/gen/literal/faceForward/2c4d14.wgsl.expected.dxc.hlsl
@@ -0,0 +1,30 @@
+void faceForward_2c4d14() {
+  float4 res = (-1.0f).xxxx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  faceForward_2c4d14();
+  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() {
+  faceForward_2c4d14();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  faceForward_2c4d14();
+  return;
+}
diff --git a/test/tint/builtins/gen/literal/faceForward/2c4d14.wgsl.expected.fxc.hlsl b/test/tint/builtins/gen/literal/faceForward/2c4d14.wgsl.expected.fxc.hlsl
new file mode 100644
index 0000000000..1109cebb9a
--- /dev/null
+++ b/test/tint/builtins/gen/literal/faceForward/2c4d14.wgsl.expected.fxc.hlsl
@@ -0,0 +1,30 @@
+void faceForward_2c4d14() {
+  float4 res = (-1.0f).xxxx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  faceForward_2c4d14();
+  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() {
+  faceForward_2c4d14();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  faceForward_2c4d14();
+  return;
+}
diff --git a/test/tint/builtins/gen/literal/faceForward/2c4d14.wgsl.expected.glsl b/test/tint/builtins/gen/literal/faceForward/2c4d14.wgsl.expected.glsl
new file mode 100644
index 0000000000..a4a1c00522
--- /dev/null
+++ b/test/tint/builtins/gen/literal/faceForward/2c4d14.wgsl.expected.glsl
@@ -0,0 +1,49 @@
+#version 310 es
+
+void faceForward_2c4d14() {
+  vec4 res = vec4(-1.0f);
+}
+
+vec4 vertex_main() {
+  faceForward_2c4d14();
+  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 faceForward_2c4d14() {
+  vec4 res = vec4(-1.0f);
+}
+
+void fragment_main() {
+  faceForward_2c4d14();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+void faceForward_2c4d14() {
+  vec4 res = vec4(-1.0f);
+}
+
+void compute_main() {
+  faceForward_2c4d14();
+}
+
+layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
+void main() {
+  compute_main();
+  return;
+}
diff --git a/test/tint/builtins/gen/literal/faceForward/2c4d14.wgsl.expected.msl b/test/tint/builtins/gen/literal/faceForward/2c4d14.wgsl.expected.msl
new file mode 100644
index 0000000000..81419e5928
--- /dev/null
+++ b/test/tint/builtins/gen/literal/faceForward/2c4d14.wgsl.expected.msl
@@ -0,0 +1,33 @@
+#include <metal_stdlib>
+
+using namespace metal;
+void faceForward_2c4d14() {
+  float4 res = float4(-1.0f);
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  faceForward_2c4d14();
+  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() {
+  faceForward_2c4d14();
+  return;
+}
+
+kernel void compute_main() {
+  faceForward_2c4d14();
+  return;
+}
+
diff --git a/test/tint/builtins/gen/literal/faceForward/2c4d14.wgsl.expected.spvasm b/test/tint/builtins/gen/literal/faceForward/2c4d14.wgsl.expected.spvasm
new file mode 100644
index 0000000000..60ade16430
--- /dev/null
+++ b/test/tint/builtins/gen/literal/faceForward/2c4d14.wgsl.expected.spvasm
@@ -0,0 +1,65 @@
+; SPIR-V
+; Version: 1.3
+; Generator: Google Tint Compiler; 0
+; Bound: 31
+; 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 %faceForward_2c4d14 "faceForward_2c4d14"
+               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
+   %float_n1 = OpConstant %float -1
+         %14 = OpConstantComposite %v4float %float_n1 %float_n1 %float_n1 %float_n1
+%_ptr_Function_v4float = OpTypePointer Function %v4float
+         %17 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%faceForward_2c4d14 = OpFunction %void None %9
+         %12 = OpLabel
+        %res = OpVariable %_ptr_Function_v4float Function %5
+               OpStore %res %14
+               OpReturn
+               OpFunctionEnd
+%vertex_main_inner = OpFunction %v4float None %17
+         %19 = OpLabel
+         %20 = OpFunctionCall %void %faceForward_2c4d14
+               OpReturnValue %5
+               OpFunctionEnd
+%vertex_main = OpFunction %void None %9
+         %22 = OpLabel
+         %23 = OpFunctionCall %v4float %vertex_main_inner
+               OpStore %value %23
+               OpStore %vertex_point_size %float_1
+               OpReturn
+               OpFunctionEnd
+%fragment_main = OpFunction %void None %9
+         %26 = OpLabel
+         %27 = OpFunctionCall %void %faceForward_2c4d14
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %29 = OpLabel
+         %30 = OpFunctionCall %void %faceForward_2c4d14
+               OpReturn
+               OpFunctionEnd
diff --git a/test/tint/builtins/gen/literal/faceForward/2c4d14.wgsl.expected.wgsl b/test/tint/builtins/gen/literal/faceForward/2c4d14.wgsl.expected.wgsl
new file mode 100644
index 0000000000..b3436ec1f1
--- /dev/null
+++ b/test/tint/builtins/gen/literal/faceForward/2c4d14.wgsl.expected.wgsl
@@ -0,0 +1,19 @@
+fn faceForward_2c4d14() {
+  var res = faceForward(vec4(1.0), vec4(1.0), vec4(1.0));
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  faceForward_2c4d14();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  faceForward_2c4d14();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  faceForward_2c4d14();
+}
diff --git a/test/tint/builtins/gen/literal/faceForward/524986.wgsl.expected.dxc.hlsl b/test/tint/builtins/gen/literal/faceForward/524986.wgsl.expected.dxc.hlsl
index d6cf67af89..f5aa2539e8 100644
--- a/test/tint/builtins/gen/literal/faceForward/524986.wgsl.expected.dxc.hlsl
+++ b/test/tint/builtins/gen/literal/faceForward/524986.wgsl.expected.dxc.hlsl
@@ -1,5 +1,5 @@
 void faceForward_524986() {
-  vector<float16_t, 3> res = faceforward((float16_t(1.0h)).xxx, (float16_t(1.0h)).xxx, (float16_t(1.0h)).xxx);
+  vector<float16_t, 3> res = (float16_t(-1.0h)).xxx;
 }
 
 struct tint_symbol {
diff --git a/test/tint/builtins/gen/literal/faceForward/524986.wgsl.expected.glsl b/test/tint/builtins/gen/literal/faceForward/524986.wgsl.expected.glsl
index 4a03ffeb0e..20c69aa42d 100644
--- a/test/tint/builtins/gen/literal/faceForward/524986.wgsl.expected.glsl
+++ b/test/tint/builtins/gen/literal/faceForward/524986.wgsl.expected.glsl
@@ -2,7 +2,7 @@
 #extension GL_AMD_gpu_shader_half_float : require
 
 void faceForward_524986() {
-  f16vec3 res = faceforward(f16vec3(1.0hf), f16vec3(1.0hf), f16vec3(1.0hf));
+  f16vec3 res = f16vec3(-1.0hf);
 }
 
 vec4 vertex_main() {
@@ -23,7 +23,7 @@ void main() {
 precision mediump float;
 
 void faceForward_524986() {
-  f16vec3 res = faceforward(f16vec3(1.0hf), f16vec3(1.0hf), f16vec3(1.0hf));
+  f16vec3 res = f16vec3(-1.0hf);
 }
 
 void fragment_main() {
@@ -38,7 +38,7 @@ void main() {
 #extension GL_AMD_gpu_shader_half_float : require
 
 void faceForward_524986() {
-  f16vec3 res = faceforward(f16vec3(1.0hf), f16vec3(1.0hf), f16vec3(1.0hf));
+  f16vec3 res = f16vec3(-1.0hf);
 }
 
 void compute_main() {
diff --git a/test/tint/builtins/gen/literal/faceForward/524986.wgsl.expected.msl b/test/tint/builtins/gen/literal/faceForward/524986.wgsl.expected.msl
index 9eea5a11ea..86bf786356 100644
--- a/test/tint/builtins/gen/literal/faceForward/524986.wgsl.expected.msl
+++ b/test/tint/builtins/gen/literal/faceForward/524986.wgsl.expected.msl
@@ -2,7 +2,7 @@
 
 using namespace metal;
 void faceForward_524986() {
-  half3 res = faceforward(half3(1.0h), half3(1.0h), half3(1.0h));
+  half3 res = half3(-1.0h);
 }
 
 struct tint_symbol {
diff --git a/test/tint/builtins/gen/literal/faceForward/524986.wgsl.expected.spvasm b/test/tint/builtins/gen/literal/faceForward/524986.wgsl.expected.spvasm
index ad5e4f5a39..ed92cbd6cb 100644
--- a/test/tint/builtins/gen/literal/faceForward/524986.wgsl.expected.spvasm
+++ b/test/tint/builtins/gen/literal/faceForward/524986.wgsl.expected.spvasm
@@ -1,14 +1,13 @@
 ; SPIR-V
 ; Version: 1.3
 ; Generator: Google Tint Compiler; 0
-; Bound: 36
+; Bound: 34
 ; Schema: 0
                OpCapability Shader
                OpCapability Float16
                OpCapability UniformAndStorageBuffer16BitAccess
                OpCapability StorageBuffer16BitAccess
                OpCapability StorageInputOutput16
-         %16 = OpExtInstImport "GLSL.std.450"
                OpMemoryModel Logical GLSL450
                OpEntryPoint Vertex %vertex_main "vertex_main" %value %vertex_point_size
                OpEntryPoint Fragment %fragment_main "fragment_main"
@@ -37,38 +36,37 @@
           %9 = OpTypeFunction %void
        %half = OpTypeFloat 16
      %v3half = OpTypeVector %half 3
-%half_0x1p_0 = OpConstant %half 0x1p+0
-         %18 = OpConstantComposite %v3half %half_0x1p_0 %half_0x1p_0 %half_0x1p_0
+%half_n0x1p_0 = OpConstant %half -0x1p+0
+         %16 = OpConstantComposite %v3half %half_n0x1p_0 %half_n0x1p_0 %half_n0x1p_0
 %_ptr_Function_v3half = OpTypePointer Function %v3half
-         %21 = OpConstantNull %v3half
-         %22 = OpTypeFunction %v4float
+         %19 = OpConstantNull %v3half
+         %20 = OpTypeFunction %v4float
     %float_1 = OpConstant %float 1
 %faceForward_524986 = OpFunction %void None %9
          %12 = OpLabel
-        %res = OpVariable %_ptr_Function_v3half Function %21
-         %13 = OpExtInst %v3half %16 FaceForward %18 %18 %18
-               OpStore %res %13
+        %res = OpVariable %_ptr_Function_v3half Function %19
+               OpStore %res %16
                OpReturn
                OpFunctionEnd
-%vertex_main_inner = OpFunction %v4float None %22
-         %24 = OpLabel
-         %25 = OpFunctionCall %void %faceForward_524986
+%vertex_main_inner = OpFunction %v4float None %20
+         %22 = OpLabel
+         %23 = OpFunctionCall %void %faceForward_524986
                OpReturnValue %5
                OpFunctionEnd
 %vertex_main = OpFunction %void None %9
-         %27 = OpLabel
-         %28 = OpFunctionCall %v4float %vertex_main_inner
-               OpStore %value %28
+         %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
-         %31 = OpLabel
-         %32 = OpFunctionCall %void %faceForward_524986
+         %29 = OpLabel
+         %30 = OpFunctionCall %void %faceForward_524986
                OpReturn
                OpFunctionEnd
 %compute_main = OpFunction %void None %9
-         %34 = OpLabel
-         %35 = OpFunctionCall %void %faceForward_524986
+         %32 = OpLabel
+         %33 = OpFunctionCall %void %faceForward_524986
                OpReturn
                OpFunctionEnd
diff --git a/test/tint/builtins/gen/literal/faceForward/5afbd5.wgsl.expected.dxc.hlsl b/test/tint/builtins/gen/literal/faceForward/5afbd5.wgsl.expected.dxc.hlsl
index c5d01ddccc..92df0efe76 100644
--- a/test/tint/builtins/gen/literal/faceForward/5afbd5.wgsl.expected.dxc.hlsl
+++ b/test/tint/builtins/gen/literal/faceForward/5afbd5.wgsl.expected.dxc.hlsl
@@ -1,5 +1,5 @@
 void faceForward_5afbd5() {
-  float3 res = faceforward((1.0f).xxx, (1.0f).xxx, (1.0f).xxx);
+  float3 res = (-1.0f).xxx;
 }
 
 struct tint_symbol {
diff --git a/test/tint/builtins/gen/literal/faceForward/5afbd5.wgsl.expected.fxc.hlsl b/test/tint/builtins/gen/literal/faceForward/5afbd5.wgsl.expected.fxc.hlsl
index c5d01ddccc..92df0efe76 100644
--- a/test/tint/builtins/gen/literal/faceForward/5afbd5.wgsl.expected.fxc.hlsl
+++ b/test/tint/builtins/gen/literal/faceForward/5afbd5.wgsl.expected.fxc.hlsl
@@ -1,5 +1,5 @@
 void faceForward_5afbd5() {
-  float3 res = faceforward((1.0f).xxx, (1.0f).xxx, (1.0f).xxx);
+  float3 res = (-1.0f).xxx;
 }
 
 struct tint_symbol {
diff --git a/test/tint/builtins/gen/literal/faceForward/5afbd5.wgsl.expected.glsl b/test/tint/builtins/gen/literal/faceForward/5afbd5.wgsl.expected.glsl
index ffec67141c..1b7f8458d2 100644
--- a/test/tint/builtins/gen/literal/faceForward/5afbd5.wgsl.expected.glsl
+++ b/test/tint/builtins/gen/literal/faceForward/5afbd5.wgsl.expected.glsl
@@ -1,7 +1,7 @@
 #version 310 es
 
 void faceForward_5afbd5() {
-  vec3 res = faceforward(vec3(1.0f), vec3(1.0f), vec3(1.0f));
+  vec3 res = vec3(-1.0f);
 }
 
 vec4 vertex_main() {
@@ -21,7 +21,7 @@ void main() {
 precision mediump float;
 
 void faceForward_5afbd5() {
-  vec3 res = faceforward(vec3(1.0f), vec3(1.0f), vec3(1.0f));
+  vec3 res = vec3(-1.0f);
 }
 
 void fragment_main() {
@@ -35,7 +35,7 @@ void main() {
 #version 310 es
 
 void faceForward_5afbd5() {
-  vec3 res = faceforward(vec3(1.0f), vec3(1.0f), vec3(1.0f));
+  vec3 res = vec3(-1.0f);
 }
 
 void compute_main() {
diff --git a/test/tint/builtins/gen/literal/faceForward/5afbd5.wgsl.expected.msl b/test/tint/builtins/gen/literal/faceForward/5afbd5.wgsl.expected.msl
index baddc993f0..d9e6c0139c 100644
--- a/test/tint/builtins/gen/literal/faceForward/5afbd5.wgsl.expected.msl
+++ b/test/tint/builtins/gen/literal/faceForward/5afbd5.wgsl.expected.msl
@@ -2,7 +2,7 @@
 
 using namespace metal;
 void faceForward_5afbd5() {
-  float3 res = faceforward(float3(1.0f), float3(1.0f), float3(1.0f));
+  float3 res = float3(-1.0f);
 }
 
 struct tint_symbol {
diff --git a/test/tint/builtins/gen/literal/faceForward/5afbd5.wgsl.expected.spvasm b/test/tint/builtins/gen/literal/faceForward/5afbd5.wgsl.expected.spvasm
index de3affd3c7..463be8927e 100644
--- a/test/tint/builtins/gen/literal/faceForward/5afbd5.wgsl.expected.spvasm
+++ b/test/tint/builtins/gen/literal/faceForward/5afbd5.wgsl.expected.spvasm
@@ -1,10 +1,9 @@
 ; SPIR-V
 ; Version: 1.3
 ; Generator: Google Tint Compiler; 0
-; Bound: 34
+; Bound: 33
 ; Schema: 0
                OpCapability Shader
-         %15 = OpExtInstImport "GLSL.std.450"
                OpMemoryModel Logical GLSL450
                OpEntryPoint Vertex %vertex_main "vertex_main" %value %vertex_point_size
                OpEntryPoint Fragment %fragment_main "fragment_main"
@@ -32,37 +31,37 @@
        %void = OpTypeVoid
           %9 = OpTypeFunction %void
     %v3float = OpTypeVector %float 3
-    %float_1 = OpConstant %float 1
-         %17 = OpConstantComposite %v3float %float_1 %float_1 %float_1
+   %float_n1 = OpConstant %float -1
+         %15 = OpConstantComposite %v3float %float_n1 %float_n1 %float_n1
 %_ptr_Function_v3float = OpTypePointer Function %v3float
-         %20 = OpConstantNull %v3float
-         %21 = OpTypeFunction %v4float
+         %18 = OpConstantNull %v3float
+         %19 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
 %faceForward_5afbd5 = OpFunction %void None %9
          %12 = OpLabel
-        %res = OpVariable %_ptr_Function_v3float Function %20
-         %13 = OpExtInst %v3float %15 FaceForward %17 %17 %17
-               OpStore %res %13
+        %res = OpVariable %_ptr_Function_v3float Function %18
+               OpStore %res %15
                OpReturn
                OpFunctionEnd
-%vertex_main_inner = OpFunction %v4float None %21
-         %23 = OpLabel
-         %24 = OpFunctionCall %void %faceForward_5afbd5
+%vertex_main_inner = OpFunction %v4float None %19
+         %21 = OpLabel
+         %22 = OpFunctionCall %void %faceForward_5afbd5
                OpReturnValue %5
                OpFunctionEnd
 %vertex_main = OpFunction %void None %9
-         %26 = OpLabel
-         %27 = OpFunctionCall %v4float %vertex_main_inner
-               OpStore %value %27
+         %24 = OpLabel
+         %25 = OpFunctionCall %v4float %vertex_main_inner
+               OpStore %value %25
                OpStore %vertex_point_size %float_1
                OpReturn
                OpFunctionEnd
 %fragment_main = OpFunction %void None %9
-         %29 = OpLabel
-         %30 = OpFunctionCall %void %faceForward_5afbd5
+         %28 = OpLabel
+         %29 = OpFunctionCall %void %faceForward_5afbd5
                OpReturn
                OpFunctionEnd
 %compute_main = OpFunction %void None %9
-         %32 = OpLabel
-         %33 = OpFunctionCall %void %faceForward_5afbd5
+         %31 = OpLabel
+         %32 = OpFunctionCall %void %faceForward_5afbd5
                OpReturn
                OpFunctionEnd
diff --git a/test/tint/builtins/gen/literal/faceForward/b316e5.wgsl.expected.dxc.hlsl b/test/tint/builtins/gen/literal/faceForward/b316e5.wgsl.expected.dxc.hlsl
index ffe25eebf2..9d9cb7027b 100644
--- a/test/tint/builtins/gen/literal/faceForward/b316e5.wgsl.expected.dxc.hlsl
+++ b/test/tint/builtins/gen/literal/faceForward/b316e5.wgsl.expected.dxc.hlsl
@@ -1,5 +1,5 @@
 void faceForward_b316e5() {
-  float4 res = faceforward((1.0f).xxxx, (1.0f).xxxx, (1.0f).xxxx);
+  float4 res = (-1.0f).xxxx;
 }
 
 struct tint_symbol {
diff --git a/test/tint/builtins/gen/literal/faceForward/b316e5.wgsl.expected.fxc.hlsl b/test/tint/builtins/gen/literal/faceForward/b316e5.wgsl.expected.fxc.hlsl
index ffe25eebf2..9d9cb7027b 100644
--- a/test/tint/builtins/gen/literal/faceForward/b316e5.wgsl.expected.fxc.hlsl
+++ b/test/tint/builtins/gen/literal/faceForward/b316e5.wgsl.expected.fxc.hlsl
@@ -1,5 +1,5 @@
 void faceForward_b316e5() {
-  float4 res = faceforward((1.0f).xxxx, (1.0f).xxxx, (1.0f).xxxx);
+  float4 res = (-1.0f).xxxx;
 }
 
 struct tint_symbol {
diff --git a/test/tint/builtins/gen/literal/faceForward/b316e5.wgsl.expected.glsl b/test/tint/builtins/gen/literal/faceForward/b316e5.wgsl.expected.glsl
index 1055f4788f..1aed1d98ab 100644
--- a/test/tint/builtins/gen/literal/faceForward/b316e5.wgsl.expected.glsl
+++ b/test/tint/builtins/gen/literal/faceForward/b316e5.wgsl.expected.glsl
@@ -1,7 +1,7 @@
 #version 310 es
 
 void faceForward_b316e5() {
-  vec4 res = faceforward(vec4(1.0f), vec4(1.0f), vec4(1.0f));
+  vec4 res = vec4(-1.0f);
 }
 
 vec4 vertex_main() {
@@ -21,7 +21,7 @@ void main() {
 precision mediump float;
 
 void faceForward_b316e5() {
-  vec4 res = faceforward(vec4(1.0f), vec4(1.0f), vec4(1.0f));
+  vec4 res = vec4(-1.0f);
 }
 
 void fragment_main() {
@@ -35,7 +35,7 @@ void main() {
 #version 310 es
 
 void faceForward_b316e5() {
-  vec4 res = faceforward(vec4(1.0f), vec4(1.0f), vec4(1.0f));
+  vec4 res = vec4(-1.0f);
 }
 
 void compute_main() {
diff --git a/test/tint/builtins/gen/literal/faceForward/b316e5.wgsl.expected.msl b/test/tint/builtins/gen/literal/faceForward/b316e5.wgsl.expected.msl
index cc1e74c8bb..3eceafa4fd 100644
--- a/test/tint/builtins/gen/literal/faceForward/b316e5.wgsl.expected.msl
+++ b/test/tint/builtins/gen/literal/faceForward/b316e5.wgsl.expected.msl
@@ -2,7 +2,7 @@
 
 using namespace metal;
 void faceForward_b316e5() {
-  float4 res = faceforward(float4(1.0f), float4(1.0f), float4(1.0f));
+  float4 res = float4(-1.0f);
 }
 
 struct tint_symbol {
diff --git a/test/tint/builtins/gen/literal/faceForward/b316e5.wgsl.expected.spvasm b/test/tint/builtins/gen/literal/faceForward/b316e5.wgsl.expected.spvasm
index ffd46e2249..e52e7e2b11 100644
--- a/test/tint/builtins/gen/literal/faceForward/b316e5.wgsl.expected.spvasm
+++ b/test/tint/builtins/gen/literal/faceForward/b316e5.wgsl.expected.spvasm
@@ -1,10 +1,9 @@
 ; SPIR-V
 ; Version: 1.3
 ; Generator: Google Tint Compiler; 0
-; Bound: 32
+; Bound: 31
 ; Schema: 0
                OpCapability Shader
-         %14 = OpExtInstImport "GLSL.std.450"
                OpMemoryModel Logical GLSL450
                OpEntryPoint Vertex %vertex_main "vertex_main" %value %vertex_point_size
                OpEntryPoint Fragment %fragment_main "fragment_main"
@@ -31,36 +30,36 @@
 %vertex_point_size = OpVariable %_ptr_Output_float Output %8
        %void = OpTypeVoid
           %9 = OpTypeFunction %void
-    %float_1 = OpConstant %float 1
-         %16 = OpConstantComposite %v4float %float_1 %float_1 %float_1 %float_1
+   %float_n1 = OpConstant %float -1
+         %14 = OpConstantComposite %v4float %float_n1 %float_n1 %float_n1 %float_n1
 %_ptr_Function_v4float = OpTypePointer Function %v4float
-         %19 = OpTypeFunction %v4float
+         %17 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
 %faceForward_b316e5 = OpFunction %void None %9
          %12 = OpLabel
         %res = OpVariable %_ptr_Function_v4float Function %5
-         %13 = OpExtInst %v4float %14 FaceForward %16 %16 %16
-               OpStore %res %13
+               OpStore %res %14
                OpReturn
                OpFunctionEnd
-%vertex_main_inner = OpFunction %v4float None %19
-         %21 = OpLabel
-         %22 = OpFunctionCall %void %faceForward_b316e5
+%vertex_main_inner = OpFunction %v4float None %17
+         %19 = OpLabel
+         %20 = OpFunctionCall %void %faceForward_b316e5
                OpReturnValue %5
                OpFunctionEnd
 %vertex_main = OpFunction %void None %9
-         %24 = OpLabel
-         %25 = OpFunctionCall %v4float %vertex_main_inner
-               OpStore %value %25
+         %22 = OpLabel
+         %23 = OpFunctionCall %v4float %vertex_main_inner
+               OpStore %value %23
                OpStore %vertex_point_size %float_1
                OpReturn
                OpFunctionEnd
 %fragment_main = OpFunction %void None %9
-         %27 = OpLabel
-         %28 = OpFunctionCall %void %faceForward_b316e5
+         %26 = OpLabel
+         %27 = OpFunctionCall %void %faceForward_b316e5
                OpReturn
                OpFunctionEnd
 %compute_main = OpFunction %void None %9
-         %30 = OpLabel
-         %31 = OpFunctionCall %void %faceForward_b316e5
+         %29 = OpLabel
+         %30 = OpFunctionCall %void %faceForward_b316e5
                OpReturn
                OpFunctionEnd
diff --git a/test/tint/builtins/gen/literal/faceForward/b42ef3.wgsl b/test/tint/builtins/gen/literal/faceForward/b42ef3.wgsl
new file mode 100644
index 0000000000..39a3905c6e
--- /dev/null
+++ b/test/tint/builtins/gen/literal/faceForward/b42ef3.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 faceForward(vec<2, fa>, vec<2, fa>, vec<2, fa>) -> vec<2, fa>
+fn faceForward_b42ef3() {
+  var res = faceForward(vec2(1.), vec2(1.), vec2(1.));
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  faceForward_b42ef3();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  faceForward_b42ef3();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  faceForward_b42ef3();
+}
diff --git a/test/tint/builtins/gen/literal/faceForward/b42ef3.wgsl.expected.dxc.hlsl b/test/tint/builtins/gen/literal/faceForward/b42ef3.wgsl.expected.dxc.hlsl
new file mode 100644
index 0000000000..a91017dedc
--- /dev/null
+++ b/test/tint/builtins/gen/literal/faceForward/b42ef3.wgsl.expected.dxc.hlsl
@@ -0,0 +1,30 @@
+void faceForward_b42ef3() {
+  float2 res = (-1.0f).xx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  faceForward_b42ef3();
+  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() {
+  faceForward_b42ef3();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  faceForward_b42ef3();
+  return;
+}
diff --git a/test/tint/builtins/gen/literal/faceForward/b42ef3.wgsl.expected.fxc.hlsl b/test/tint/builtins/gen/literal/faceForward/b42ef3.wgsl.expected.fxc.hlsl
new file mode 100644
index 0000000000..a91017dedc
--- /dev/null
+++ b/test/tint/builtins/gen/literal/faceForward/b42ef3.wgsl.expected.fxc.hlsl
@@ -0,0 +1,30 @@
+void faceForward_b42ef3() {
+  float2 res = (-1.0f).xx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  faceForward_b42ef3();
+  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() {
+  faceForward_b42ef3();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  faceForward_b42ef3();
+  return;
+}
diff --git a/test/tint/builtins/gen/literal/faceForward/b42ef3.wgsl.expected.glsl b/test/tint/builtins/gen/literal/faceForward/b42ef3.wgsl.expected.glsl
new file mode 100644
index 0000000000..316c6ff1b4
--- /dev/null
+++ b/test/tint/builtins/gen/literal/faceForward/b42ef3.wgsl.expected.glsl
@@ -0,0 +1,49 @@
+#version 310 es
+
+void faceForward_b42ef3() {
+  vec2 res = vec2(-1.0f);
+}
+
+vec4 vertex_main() {
+  faceForward_b42ef3();
+  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 faceForward_b42ef3() {
+  vec2 res = vec2(-1.0f);
+}
+
+void fragment_main() {
+  faceForward_b42ef3();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+void faceForward_b42ef3() {
+  vec2 res = vec2(-1.0f);
+}
+
+void compute_main() {
+  faceForward_b42ef3();
+}
+
+layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
+void main() {
+  compute_main();
+  return;
+}
diff --git a/test/tint/builtins/gen/literal/faceForward/b42ef3.wgsl.expected.msl b/test/tint/builtins/gen/literal/faceForward/b42ef3.wgsl.expected.msl
new file mode 100644
index 0000000000..6a9646f3ad
--- /dev/null
+++ b/test/tint/builtins/gen/literal/faceForward/b42ef3.wgsl.expected.msl
@@ -0,0 +1,33 @@
+#include <metal_stdlib>
+
+using namespace metal;
+void faceForward_b42ef3() {
+  float2 res = float2(-1.0f);
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  faceForward_b42ef3();
+  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() {
+  faceForward_b42ef3();
+  return;
+}
+
+kernel void compute_main() {
+  faceForward_b42ef3();
+  return;
+}
+
diff --git a/test/tint/builtins/gen/literal/faceForward/b42ef3.wgsl.expected.spvasm b/test/tint/builtins/gen/literal/faceForward/b42ef3.wgsl.expected.spvasm
new file mode 100644
index 0000000000..63df79352f
--- /dev/null
+++ b/test/tint/builtins/gen/literal/faceForward/b42ef3.wgsl.expected.spvasm
@@ -0,0 +1,67 @@
+; SPIR-V
+; Version: 1.3
+; Generator: Google Tint Compiler; 0
+; Bound: 33
+; 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 %faceForward_b42ef3 "faceForward_b42ef3"
+               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
+    %v2float = OpTypeVector %float 2
+   %float_n1 = OpConstant %float -1
+         %15 = OpConstantComposite %v2float %float_n1 %float_n1
+%_ptr_Function_v2float = OpTypePointer Function %v2float
+         %18 = OpConstantNull %v2float
+         %19 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%faceForward_b42ef3 = OpFunction %void None %9
+         %12 = OpLabel
+        %res = OpVariable %_ptr_Function_v2float Function %18
+               OpStore %res %15
+               OpReturn
+               OpFunctionEnd
+%vertex_main_inner = OpFunction %v4float None %19
+         %21 = OpLabel
+         %22 = OpFunctionCall %void %faceForward_b42ef3
+               OpReturnValue %5
+               OpFunctionEnd
+%vertex_main = OpFunction %void None %9
+         %24 = OpLabel
+         %25 = OpFunctionCall %v4float %vertex_main_inner
+               OpStore %value %25
+               OpStore %vertex_point_size %float_1
+               OpReturn
+               OpFunctionEnd
+%fragment_main = OpFunction %void None %9
+         %28 = OpLabel
+         %29 = OpFunctionCall %void %faceForward_b42ef3
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %31 = OpLabel
+         %32 = OpFunctionCall %void %faceForward_b42ef3
+               OpReturn
+               OpFunctionEnd
diff --git a/test/tint/builtins/gen/literal/faceForward/b42ef3.wgsl.expected.wgsl b/test/tint/builtins/gen/literal/faceForward/b42ef3.wgsl.expected.wgsl
new file mode 100644
index 0000000000..ffa7ab5000
--- /dev/null
+++ b/test/tint/builtins/gen/literal/faceForward/b42ef3.wgsl.expected.wgsl
@@ -0,0 +1,19 @@
+fn faceForward_b42ef3() {
+  var res = faceForward(vec2(1.0), vec2(1.0), vec2(1.0));
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  faceForward_b42ef3();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  faceForward_b42ef3();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  faceForward_b42ef3();
+}
diff --git a/test/tint/builtins/gen/literal/faceForward/cc63dc.wgsl.expected.dxc.hlsl b/test/tint/builtins/gen/literal/faceForward/cc63dc.wgsl.expected.dxc.hlsl
index ec3d4f84bc..59991b58c6 100644
--- a/test/tint/builtins/gen/literal/faceForward/cc63dc.wgsl.expected.dxc.hlsl
+++ b/test/tint/builtins/gen/literal/faceForward/cc63dc.wgsl.expected.dxc.hlsl
@@ -1,5 +1,5 @@
 void faceForward_cc63dc() {
-  vector<float16_t, 4> res = faceforward((float16_t(1.0h)).xxxx, (float16_t(1.0h)).xxxx, (float16_t(1.0h)).xxxx);
+  vector<float16_t, 4> res = (float16_t(-1.0h)).xxxx;
 }
 
 struct tint_symbol {
diff --git a/test/tint/builtins/gen/literal/faceForward/cc63dc.wgsl.expected.glsl b/test/tint/builtins/gen/literal/faceForward/cc63dc.wgsl.expected.glsl
index 6a53621707..aec75dcb60 100644
--- a/test/tint/builtins/gen/literal/faceForward/cc63dc.wgsl.expected.glsl
+++ b/test/tint/builtins/gen/literal/faceForward/cc63dc.wgsl.expected.glsl
@@ -2,7 +2,7 @@
 #extension GL_AMD_gpu_shader_half_float : require
 
 void faceForward_cc63dc() {
-  f16vec4 res = faceforward(f16vec4(1.0hf), f16vec4(1.0hf), f16vec4(1.0hf));
+  f16vec4 res = f16vec4(-1.0hf);
 }
 
 vec4 vertex_main() {
@@ -23,7 +23,7 @@ void main() {
 precision mediump float;
 
 void faceForward_cc63dc() {
-  f16vec4 res = faceforward(f16vec4(1.0hf), f16vec4(1.0hf), f16vec4(1.0hf));
+  f16vec4 res = f16vec4(-1.0hf);
 }
 
 void fragment_main() {
@@ -38,7 +38,7 @@ void main() {
 #extension GL_AMD_gpu_shader_half_float : require
 
 void faceForward_cc63dc() {
-  f16vec4 res = faceforward(f16vec4(1.0hf), f16vec4(1.0hf), f16vec4(1.0hf));
+  f16vec4 res = f16vec4(-1.0hf);
 }
 
 void compute_main() {
diff --git a/test/tint/builtins/gen/literal/faceForward/cc63dc.wgsl.expected.msl b/test/tint/builtins/gen/literal/faceForward/cc63dc.wgsl.expected.msl
index 423f02e55b..f3e2791ae5 100644
--- a/test/tint/builtins/gen/literal/faceForward/cc63dc.wgsl.expected.msl
+++ b/test/tint/builtins/gen/literal/faceForward/cc63dc.wgsl.expected.msl
@@ -2,7 +2,7 @@
 
 using namespace metal;
 void faceForward_cc63dc() {
-  half4 res = faceforward(half4(1.0h), half4(1.0h), half4(1.0h));
+  half4 res = half4(-1.0h);
 }
 
 struct tint_symbol {
diff --git a/test/tint/builtins/gen/literal/faceForward/cc63dc.wgsl.expected.spvasm b/test/tint/builtins/gen/literal/faceForward/cc63dc.wgsl.expected.spvasm
index a5052c50d5..a9dc6f1cc5 100644
--- a/test/tint/builtins/gen/literal/faceForward/cc63dc.wgsl.expected.spvasm
+++ b/test/tint/builtins/gen/literal/faceForward/cc63dc.wgsl.expected.spvasm
@@ -1,14 +1,13 @@
 ; SPIR-V
 ; Version: 1.3
 ; Generator: Google Tint Compiler; 0
-; Bound: 36
+; Bound: 34
 ; Schema: 0
                OpCapability Shader
                OpCapability Float16
                OpCapability UniformAndStorageBuffer16BitAccess
                OpCapability StorageBuffer16BitAccess
                OpCapability StorageInputOutput16
-         %16 = OpExtInstImport "GLSL.std.450"
                OpMemoryModel Logical GLSL450
                OpEntryPoint Vertex %vertex_main "vertex_main" %value %vertex_point_size
                OpEntryPoint Fragment %fragment_main "fragment_main"
@@ -37,38 +36,37 @@
           %9 = OpTypeFunction %void
        %half = OpTypeFloat 16
      %v4half = OpTypeVector %half 4
-%half_0x1p_0 = OpConstant %half 0x1p+0
-         %18 = OpConstantComposite %v4half %half_0x1p_0 %half_0x1p_0 %half_0x1p_0 %half_0x1p_0
+%half_n0x1p_0 = OpConstant %half -0x1p+0
+         %16 = OpConstantComposite %v4half %half_n0x1p_0 %half_n0x1p_0 %half_n0x1p_0 %half_n0x1p_0
 %_ptr_Function_v4half = OpTypePointer Function %v4half
-         %21 = OpConstantNull %v4half
-         %22 = OpTypeFunction %v4float
+         %19 = OpConstantNull %v4half
+         %20 = OpTypeFunction %v4float
     %float_1 = OpConstant %float 1
 %faceForward_cc63dc = OpFunction %void None %9
          %12 = OpLabel
-        %res = OpVariable %_ptr_Function_v4half Function %21
-         %13 = OpExtInst %v4half %16 FaceForward %18 %18 %18
-               OpStore %res %13
+        %res = OpVariable %_ptr_Function_v4half Function %19
+               OpStore %res %16
                OpReturn
                OpFunctionEnd
-%vertex_main_inner = OpFunction %v4float None %22
-         %24 = OpLabel
-         %25 = OpFunctionCall %void %faceForward_cc63dc
+%vertex_main_inner = OpFunction %v4float None %20
+         %22 = OpLabel
+         %23 = OpFunctionCall %void %faceForward_cc63dc
                OpReturnValue %5
                OpFunctionEnd
 %vertex_main = OpFunction %void None %9
-         %27 = OpLabel
-         %28 = OpFunctionCall %v4float %vertex_main_inner
-               OpStore %value %28
+         %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
-         %31 = OpLabel
-         %32 = OpFunctionCall %void %faceForward_cc63dc
+         %29 = OpLabel
+         %30 = OpFunctionCall %void %faceForward_cc63dc
                OpReturn
                OpFunctionEnd
 %compute_main = OpFunction %void None %9
-         %34 = OpLabel
-         %35 = OpFunctionCall %void %faceForward_cc63dc
+         %32 = OpLabel
+         %33 = OpFunctionCall %void %faceForward_cc63dc
                OpReturn
                OpFunctionEnd
diff --git a/test/tint/builtins/gen/literal/faceForward/e6908b.wgsl.expected.dxc.hlsl b/test/tint/builtins/gen/literal/faceForward/e6908b.wgsl.expected.dxc.hlsl
index 82a14e79f9..a69c897f9d 100644
--- a/test/tint/builtins/gen/literal/faceForward/e6908b.wgsl.expected.dxc.hlsl
+++ b/test/tint/builtins/gen/literal/faceForward/e6908b.wgsl.expected.dxc.hlsl
@@ -1,5 +1,5 @@
 void faceForward_e6908b() {
-  float2 res = faceforward((1.0f).xx, (1.0f).xx, (1.0f).xx);
+  float2 res = (-1.0f).xx;
 }
 
 struct tint_symbol {
diff --git a/test/tint/builtins/gen/literal/faceForward/e6908b.wgsl.expected.fxc.hlsl b/test/tint/builtins/gen/literal/faceForward/e6908b.wgsl.expected.fxc.hlsl
index 82a14e79f9..a69c897f9d 100644
--- a/test/tint/builtins/gen/literal/faceForward/e6908b.wgsl.expected.fxc.hlsl
+++ b/test/tint/builtins/gen/literal/faceForward/e6908b.wgsl.expected.fxc.hlsl
@@ -1,5 +1,5 @@
 void faceForward_e6908b() {
-  float2 res = faceforward((1.0f).xx, (1.0f).xx, (1.0f).xx);
+  float2 res = (-1.0f).xx;
 }
 
 struct tint_symbol {
diff --git a/test/tint/builtins/gen/literal/faceForward/e6908b.wgsl.expected.glsl b/test/tint/builtins/gen/literal/faceForward/e6908b.wgsl.expected.glsl
index 78a59cca7b..274e8f866c 100644
--- a/test/tint/builtins/gen/literal/faceForward/e6908b.wgsl.expected.glsl
+++ b/test/tint/builtins/gen/literal/faceForward/e6908b.wgsl.expected.glsl
@@ -1,7 +1,7 @@
 #version 310 es
 
 void faceForward_e6908b() {
-  vec2 res = faceforward(vec2(1.0f), vec2(1.0f), vec2(1.0f));
+  vec2 res = vec2(-1.0f);
 }
 
 vec4 vertex_main() {
@@ -21,7 +21,7 @@ void main() {
 precision mediump float;
 
 void faceForward_e6908b() {
-  vec2 res = faceforward(vec2(1.0f), vec2(1.0f), vec2(1.0f));
+  vec2 res = vec2(-1.0f);
 }
 
 void fragment_main() {
@@ -35,7 +35,7 @@ void main() {
 #version 310 es
 
 void faceForward_e6908b() {
-  vec2 res = faceforward(vec2(1.0f), vec2(1.0f), vec2(1.0f));
+  vec2 res = vec2(-1.0f);
 }
 
 void compute_main() {
diff --git a/test/tint/builtins/gen/literal/faceForward/e6908b.wgsl.expected.msl b/test/tint/builtins/gen/literal/faceForward/e6908b.wgsl.expected.msl
index db005d9004..026c5486ba 100644
--- a/test/tint/builtins/gen/literal/faceForward/e6908b.wgsl.expected.msl
+++ b/test/tint/builtins/gen/literal/faceForward/e6908b.wgsl.expected.msl
@@ -2,7 +2,7 @@
 
 using namespace metal;
 void faceForward_e6908b() {
-  float2 res = faceforward(float2(1.0f), float2(1.0f), float2(1.0f));
+  float2 res = float2(-1.0f);
 }
 
 struct tint_symbol {
diff --git a/test/tint/builtins/gen/literal/faceForward/e6908b.wgsl.expected.spvasm b/test/tint/builtins/gen/literal/faceForward/e6908b.wgsl.expected.spvasm
index 8e1293d69d..c759014083 100644
--- a/test/tint/builtins/gen/literal/faceForward/e6908b.wgsl.expected.spvasm
+++ b/test/tint/builtins/gen/literal/faceForward/e6908b.wgsl.expected.spvasm
@@ -1,10 +1,9 @@
 ; SPIR-V
 ; Version: 1.3
 ; Generator: Google Tint Compiler; 0
-; Bound: 34
+; Bound: 33
 ; Schema: 0
                OpCapability Shader
-         %15 = OpExtInstImport "GLSL.std.450"
                OpMemoryModel Logical GLSL450
                OpEntryPoint Vertex %vertex_main "vertex_main" %value %vertex_point_size
                OpEntryPoint Fragment %fragment_main "fragment_main"
@@ -32,37 +31,37 @@
        %void = OpTypeVoid
           %9 = OpTypeFunction %void
     %v2float = OpTypeVector %float 2
-    %float_1 = OpConstant %float 1
-         %17 = OpConstantComposite %v2float %float_1 %float_1
+   %float_n1 = OpConstant %float -1
+         %15 = OpConstantComposite %v2float %float_n1 %float_n1
 %_ptr_Function_v2float = OpTypePointer Function %v2float
-         %20 = OpConstantNull %v2float
-         %21 = OpTypeFunction %v4float
+         %18 = OpConstantNull %v2float
+         %19 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
 %faceForward_e6908b = OpFunction %void None %9
          %12 = OpLabel
-        %res = OpVariable %_ptr_Function_v2float Function %20
-         %13 = OpExtInst %v2float %15 FaceForward %17 %17 %17
-               OpStore %res %13
+        %res = OpVariable %_ptr_Function_v2float Function %18
+               OpStore %res %15
                OpReturn
                OpFunctionEnd
-%vertex_main_inner = OpFunction %v4float None %21
-         %23 = OpLabel
-         %24 = OpFunctionCall %void %faceForward_e6908b
+%vertex_main_inner = OpFunction %v4float None %19
+         %21 = OpLabel
+         %22 = OpFunctionCall %void %faceForward_e6908b
                OpReturnValue %5
                OpFunctionEnd
 %vertex_main = OpFunction %void None %9
-         %26 = OpLabel
-         %27 = OpFunctionCall %v4float %vertex_main_inner
-               OpStore %value %27
+         %24 = OpLabel
+         %25 = OpFunctionCall %v4float %vertex_main_inner
+               OpStore %value %25
                OpStore %vertex_point_size %float_1
                OpReturn
                OpFunctionEnd
 %fragment_main = OpFunction %void None %9
-         %29 = OpLabel
-         %30 = OpFunctionCall %void %faceForward_e6908b
+         %28 = OpLabel
+         %29 = OpFunctionCall %void %faceForward_e6908b
                OpReturn
                OpFunctionEnd
 %compute_main = OpFunction %void None %9
-         %32 = OpLabel
-         %33 = OpFunctionCall %void %faceForward_e6908b
+         %31 = OpLabel
+         %32 = OpFunctionCall %void %faceForward_e6908b
                OpReturn
                OpFunctionEnd
diff --git a/test/tint/builtins/gen/literal/faceForward/fb0f2e.wgsl.expected.dxc.hlsl b/test/tint/builtins/gen/literal/faceForward/fb0f2e.wgsl.expected.dxc.hlsl
index 0cea4d4e6b..b8a27dc185 100644
--- a/test/tint/builtins/gen/literal/faceForward/fb0f2e.wgsl.expected.dxc.hlsl
+++ b/test/tint/builtins/gen/literal/faceForward/fb0f2e.wgsl.expected.dxc.hlsl
@@ -1,5 +1,5 @@
 void faceForward_fb0f2e() {
-  vector<float16_t, 2> res = faceforward((float16_t(1.0h)).xx, (float16_t(1.0h)).xx, (float16_t(1.0h)).xx);
+  vector<float16_t, 2> res = (float16_t(-1.0h)).xx;
 }
 
 struct tint_symbol {
diff --git a/test/tint/builtins/gen/literal/faceForward/fb0f2e.wgsl.expected.glsl b/test/tint/builtins/gen/literal/faceForward/fb0f2e.wgsl.expected.glsl
index 91747bf50d..c5385adda1 100644
--- a/test/tint/builtins/gen/literal/faceForward/fb0f2e.wgsl.expected.glsl
+++ b/test/tint/builtins/gen/literal/faceForward/fb0f2e.wgsl.expected.glsl
@@ -2,7 +2,7 @@
 #extension GL_AMD_gpu_shader_half_float : require
 
 void faceForward_fb0f2e() {
-  f16vec2 res = faceforward(f16vec2(1.0hf), f16vec2(1.0hf), f16vec2(1.0hf));
+  f16vec2 res = f16vec2(-1.0hf);
 }
 
 vec4 vertex_main() {
@@ -23,7 +23,7 @@ void main() {
 precision mediump float;
 
 void faceForward_fb0f2e() {
-  f16vec2 res = faceforward(f16vec2(1.0hf), f16vec2(1.0hf), f16vec2(1.0hf));
+  f16vec2 res = f16vec2(-1.0hf);
 }
 
 void fragment_main() {
@@ -38,7 +38,7 @@ void main() {
 #extension GL_AMD_gpu_shader_half_float : require
 
 void faceForward_fb0f2e() {
-  f16vec2 res = faceforward(f16vec2(1.0hf), f16vec2(1.0hf), f16vec2(1.0hf));
+  f16vec2 res = f16vec2(-1.0hf);
 }
 
 void compute_main() {
diff --git a/test/tint/builtins/gen/literal/faceForward/fb0f2e.wgsl.expected.msl b/test/tint/builtins/gen/literal/faceForward/fb0f2e.wgsl.expected.msl
index ef0dd6db54..91409872de 100644
--- a/test/tint/builtins/gen/literal/faceForward/fb0f2e.wgsl.expected.msl
+++ b/test/tint/builtins/gen/literal/faceForward/fb0f2e.wgsl.expected.msl
@@ -2,7 +2,7 @@
 
 using namespace metal;
 void faceForward_fb0f2e() {
-  half2 res = faceforward(half2(1.0h), half2(1.0h), half2(1.0h));
+  half2 res = half2(-1.0h);
 }
 
 struct tint_symbol {
diff --git a/test/tint/builtins/gen/literal/faceForward/fb0f2e.wgsl.expected.spvasm b/test/tint/builtins/gen/literal/faceForward/fb0f2e.wgsl.expected.spvasm
index b3e1ff05af..7d7f40b8c5 100644
--- a/test/tint/builtins/gen/literal/faceForward/fb0f2e.wgsl.expected.spvasm
+++ b/test/tint/builtins/gen/literal/faceForward/fb0f2e.wgsl.expected.spvasm
@@ -1,14 +1,13 @@
 ; SPIR-V
 ; Version: 1.3
 ; Generator: Google Tint Compiler; 0
-; Bound: 36
+; Bound: 34
 ; Schema: 0
                OpCapability Shader
                OpCapability Float16
                OpCapability UniformAndStorageBuffer16BitAccess
                OpCapability StorageBuffer16BitAccess
                OpCapability StorageInputOutput16
-         %16 = OpExtInstImport "GLSL.std.450"
                OpMemoryModel Logical GLSL450
                OpEntryPoint Vertex %vertex_main "vertex_main" %value %vertex_point_size
                OpEntryPoint Fragment %fragment_main "fragment_main"
@@ -37,38 +36,37 @@
           %9 = OpTypeFunction %void
        %half = OpTypeFloat 16
      %v2half = OpTypeVector %half 2
-%half_0x1p_0 = OpConstant %half 0x1p+0
-         %18 = OpConstantComposite %v2half %half_0x1p_0 %half_0x1p_0
+%half_n0x1p_0 = OpConstant %half -0x1p+0
+         %16 = OpConstantComposite %v2half %half_n0x1p_0 %half_n0x1p_0
 %_ptr_Function_v2half = OpTypePointer Function %v2half
-         %21 = OpConstantNull %v2half
-         %22 = OpTypeFunction %v4float
+         %19 = OpConstantNull %v2half
+         %20 = OpTypeFunction %v4float
     %float_1 = OpConstant %float 1
 %faceForward_fb0f2e = OpFunction %void None %9
          %12 = OpLabel
-        %res = OpVariable %_ptr_Function_v2half Function %21
-         %13 = OpExtInst %v2half %16 FaceForward %18 %18 %18
-               OpStore %res %13
+        %res = OpVariable %_ptr_Function_v2half Function %19
+               OpStore %res %16
                OpReturn
                OpFunctionEnd
-%vertex_main_inner = OpFunction %v4float None %22
-         %24 = OpLabel
-         %25 = OpFunctionCall %void %faceForward_fb0f2e
+%vertex_main_inner = OpFunction %v4float None %20
+         %22 = OpLabel
+         %23 = OpFunctionCall %void %faceForward_fb0f2e
                OpReturnValue %5
                OpFunctionEnd
 %vertex_main = OpFunction %void None %9
-         %27 = OpLabel
-         %28 = OpFunctionCall %v4float %vertex_main_inner
-               OpStore %value %28
+         %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
-         %31 = OpLabel
-         %32 = OpFunctionCall %void %faceForward_fb0f2e
+         %29 = OpLabel
+         %30 = OpFunctionCall %void %faceForward_fb0f2e
                OpReturn
                OpFunctionEnd
 %compute_main = OpFunction %void None %9
-         %34 = OpLabel
-         %35 = OpFunctionCall %void %faceForward_fb0f2e
+         %32 = OpLabel
+         %33 = OpFunctionCall %void %faceForward_fb0f2e
                OpReturn
                OpFunctionEnd
diff --git a/test/tint/builtins/gen/literal/faceForward/fe522b.wgsl b/test/tint/builtins/gen/literal/faceForward/fe522b.wgsl
new file mode 100644
index 0000000000..fe8be541cb
--- /dev/null
+++ b/test/tint/builtins/gen/literal/faceForward/fe522b.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 faceForward(vec<3, fa>, vec<3, fa>, vec<3, fa>) -> vec<3, fa>
+fn faceForward_fe522b() {
+  var res = faceForward(vec3(1.), vec3(1.), vec3(1.));
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  faceForward_fe522b();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  faceForward_fe522b();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  faceForward_fe522b();
+}
diff --git a/test/tint/builtins/gen/literal/faceForward/fe522b.wgsl.expected.dxc.hlsl b/test/tint/builtins/gen/literal/faceForward/fe522b.wgsl.expected.dxc.hlsl
new file mode 100644
index 0000000000..706394b401
--- /dev/null
+++ b/test/tint/builtins/gen/literal/faceForward/fe522b.wgsl.expected.dxc.hlsl
@@ -0,0 +1,30 @@
+void faceForward_fe522b() {
+  float3 res = (-1.0f).xxx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  faceForward_fe522b();
+  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() {
+  faceForward_fe522b();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  faceForward_fe522b();
+  return;
+}
diff --git a/test/tint/builtins/gen/literal/faceForward/fe522b.wgsl.expected.fxc.hlsl b/test/tint/builtins/gen/literal/faceForward/fe522b.wgsl.expected.fxc.hlsl
new file mode 100644
index 0000000000..706394b401
--- /dev/null
+++ b/test/tint/builtins/gen/literal/faceForward/fe522b.wgsl.expected.fxc.hlsl
@@ -0,0 +1,30 @@
+void faceForward_fe522b() {
+  float3 res = (-1.0f).xxx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  faceForward_fe522b();
+  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() {
+  faceForward_fe522b();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  faceForward_fe522b();
+  return;
+}
diff --git a/test/tint/builtins/gen/literal/faceForward/fe522b.wgsl.expected.glsl b/test/tint/builtins/gen/literal/faceForward/fe522b.wgsl.expected.glsl
new file mode 100644
index 0000000000..3e26400a49
--- /dev/null
+++ b/test/tint/builtins/gen/literal/faceForward/fe522b.wgsl.expected.glsl
@@ -0,0 +1,49 @@
+#version 310 es
+
+void faceForward_fe522b() {
+  vec3 res = vec3(-1.0f);
+}
+
+vec4 vertex_main() {
+  faceForward_fe522b();
+  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 faceForward_fe522b() {
+  vec3 res = vec3(-1.0f);
+}
+
+void fragment_main() {
+  faceForward_fe522b();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+void faceForward_fe522b() {
+  vec3 res = vec3(-1.0f);
+}
+
+void compute_main() {
+  faceForward_fe522b();
+}
+
+layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
+void main() {
+  compute_main();
+  return;
+}
diff --git a/test/tint/builtins/gen/literal/faceForward/fe522b.wgsl.expected.msl b/test/tint/builtins/gen/literal/faceForward/fe522b.wgsl.expected.msl
new file mode 100644
index 0000000000..4368ad5048
--- /dev/null
+++ b/test/tint/builtins/gen/literal/faceForward/fe522b.wgsl.expected.msl
@@ -0,0 +1,33 @@
+#include <metal_stdlib>
+
+using namespace metal;
+void faceForward_fe522b() {
+  float3 res = float3(-1.0f);
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  faceForward_fe522b();
+  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() {
+  faceForward_fe522b();
+  return;
+}
+
+kernel void compute_main() {
+  faceForward_fe522b();
+  return;
+}
+
diff --git a/test/tint/builtins/gen/literal/faceForward/fe522b.wgsl.expected.spvasm b/test/tint/builtins/gen/literal/faceForward/fe522b.wgsl.expected.spvasm
new file mode 100644
index 0000000000..d3a5b96ea2
--- /dev/null
+++ b/test/tint/builtins/gen/literal/faceForward/fe522b.wgsl.expected.spvasm
@@ -0,0 +1,67 @@
+; SPIR-V
+; Version: 1.3
+; Generator: Google Tint Compiler; 0
+; Bound: 33
+; 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 %faceForward_fe522b "faceForward_fe522b"
+               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
+    %v3float = OpTypeVector %float 3
+   %float_n1 = OpConstant %float -1
+         %15 = OpConstantComposite %v3float %float_n1 %float_n1 %float_n1
+%_ptr_Function_v3float = OpTypePointer Function %v3float
+         %18 = OpConstantNull %v3float
+         %19 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%faceForward_fe522b = OpFunction %void None %9
+         %12 = OpLabel
+        %res = OpVariable %_ptr_Function_v3float Function %18
+               OpStore %res %15
+               OpReturn
+               OpFunctionEnd
+%vertex_main_inner = OpFunction %v4float None %19
+         %21 = OpLabel
+         %22 = OpFunctionCall %void %faceForward_fe522b
+               OpReturnValue %5
+               OpFunctionEnd
+%vertex_main = OpFunction %void None %9
+         %24 = OpLabel
+         %25 = OpFunctionCall %v4float %vertex_main_inner
+               OpStore %value %25
+               OpStore %vertex_point_size %float_1
+               OpReturn
+               OpFunctionEnd
+%fragment_main = OpFunction %void None %9
+         %28 = OpLabel
+         %29 = OpFunctionCall %void %faceForward_fe522b
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %31 = OpLabel
+         %32 = OpFunctionCall %void %faceForward_fe522b
+               OpReturn
+               OpFunctionEnd
diff --git a/test/tint/builtins/gen/literal/faceForward/fe522b.wgsl.expected.wgsl b/test/tint/builtins/gen/literal/faceForward/fe522b.wgsl.expected.wgsl
new file mode 100644
index 0000000000..bf51849a3d
--- /dev/null
+++ b/test/tint/builtins/gen/literal/faceForward/fe522b.wgsl.expected.wgsl
@@ -0,0 +1,19 @@
+fn faceForward_fe522b() {
+  var res = faceForward(vec3(1.0), vec3(1.0), vec3(1.0));
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  faceForward_fe522b();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  faceForward_fe522b();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  faceForward_fe522b();
+}
diff --git a/test/tint/builtins/gen/var/faceForward/2c4d14.wgsl b/test/tint/builtins/gen/var/faceForward/2c4d14.wgsl
new file mode 100644
index 0000000000..40b85e1b23
--- /dev/null
+++ b/test/tint/builtins/gen/var/faceForward/2c4d14.wgsl
@@ -0,0 +1,46 @@
+// 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 faceForward(vec<4, fa>, vec<4, fa>, vec<4, fa>) -> vec<4, fa>
+fn faceForward_2c4d14() {
+  const arg_0 = vec4(1.);
+  const arg_1 = vec4(1.);
+  const arg_2 = vec4(1.);
+  var res = faceForward(arg_0, arg_1, arg_2);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  faceForward_2c4d14();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  faceForward_2c4d14();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  faceForward_2c4d14();
+}
diff --git a/test/tint/builtins/gen/var/faceForward/2c4d14.wgsl.expected.dxc.hlsl b/test/tint/builtins/gen/var/faceForward/2c4d14.wgsl.expected.dxc.hlsl
new file mode 100644
index 0000000000..1109cebb9a
--- /dev/null
+++ b/test/tint/builtins/gen/var/faceForward/2c4d14.wgsl.expected.dxc.hlsl
@@ -0,0 +1,30 @@
+void faceForward_2c4d14() {
+  float4 res = (-1.0f).xxxx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  faceForward_2c4d14();
+  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() {
+  faceForward_2c4d14();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  faceForward_2c4d14();
+  return;
+}
diff --git a/test/tint/builtins/gen/var/faceForward/2c4d14.wgsl.expected.fxc.hlsl b/test/tint/builtins/gen/var/faceForward/2c4d14.wgsl.expected.fxc.hlsl
new file mode 100644
index 0000000000..1109cebb9a
--- /dev/null
+++ b/test/tint/builtins/gen/var/faceForward/2c4d14.wgsl.expected.fxc.hlsl
@@ -0,0 +1,30 @@
+void faceForward_2c4d14() {
+  float4 res = (-1.0f).xxxx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  faceForward_2c4d14();
+  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() {
+  faceForward_2c4d14();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  faceForward_2c4d14();
+  return;
+}
diff --git a/test/tint/builtins/gen/var/faceForward/2c4d14.wgsl.expected.glsl b/test/tint/builtins/gen/var/faceForward/2c4d14.wgsl.expected.glsl
new file mode 100644
index 0000000000..a4a1c00522
--- /dev/null
+++ b/test/tint/builtins/gen/var/faceForward/2c4d14.wgsl.expected.glsl
@@ -0,0 +1,49 @@
+#version 310 es
+
+void faceForward_2c4d14() {
+  vec4 res = vec4(-1.0f);
+}
+
+vec4 vertex_main() {
+  faceForward_2c4d14();
+  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 faceForward_2c4d14() {
+  vec4 res = vec4(-1.0f);
+}
+
+void fragment_main() {
+  faceForward_2c4d14();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+void faceForward_2c4d14() {
+  vec4 res = vec4(-1.0f);
+}
+
+void compute_main() {
+  faceForward_2c4d14();
+}
+
+layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
+void main() {
+  compute_main();
+  return;
+}
diff --git a/test/tint/builtins/gen/var/faceForward/2c4d14.wgsl.expected.msl b/test/tint/builtins/gen/var/faceForward/2c4d14.wgsl.expected.msl
new file mode 100644
index 0000000000..81419e5928
--- /dev/null
+++ b/test/tint/builtins/gen/var/faceForward/2c4d14.wgsl.expected.msl
@@ -0,0 +1,33 @@
+#include <metal_stdlib>
+
+using namespace metal;
+void faceForward_2c4d14() {
+  float4 res = float4(-1.0f);
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  faceForward_2c4d14();
+  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() {
+  faceForward_2c4d14();
+  return;
+}
+
+kernel void compute_main() {
+  faceForward_2c4d14();
+  return;
+}
+
diff --git a/test/tint/builtins/gen/var/faceForward/2c4d14.wgsl.expected.spvasm b/test/tint/builtins/gen/var/faceForward/2c4d14.wgsl.expected.spvasm
new file mode 100644
index 0000000000..60ade16430
--- /dev/null
+++ b/test/tint/builtins/gen/var/faceForward/2c4d14.wgsl.expected.spvasm
@@ -0,0 +1,65 @@
+; SPIR-V
+; Version: 1.3
+; Generator: Google Tint Compiler; 0
+; Bound: 31
+; 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 %faceForward_2c4d14 "faceForward_2c4d14"
+               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
+   %float_n1 = OpConstant %float -1
+         %14 = OpConstantComposite %v4float %float_n1 %float_n1 %float_n1 %float_n1
+%_ptr_Function_v4float = OpTypePointer Function %v4float
+         %17 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%faceForward_2c4d14 = OpFunction %void None %9
+         %12 = OpLabel
+        %res = OpVariable %_ptr_Function_v4float Function %5
+               OpStore %res %14
+               OpReturn
+               OpFunctionEnd
+%vertex_main_inner = OpFunction %v4float None %17
+         %19 = OpLabel
+         %20 = OpFunctionCall %void %faceForward_2c4d14
+               OpReturnValue %5
+               OpFunctionEnd
+%vertex_main = OpFunction %void None %9
+         %22 = OpLabel
+         %23 = OpFunctionCall %v4float %vertex_main_inner
+               OpStore %value %23
+               OpStore %vertex_point_size %float_1
+               OpReturn
+               OpFunctionEnd
+%fragment_main = OpFunction %void None %9
+         %26 = OpLabel
+         %27 = OpFunctionCall %void %faceForward_2c4d14
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %29 = OpLabel
+         %30 = OpFunctionCall %void %faceForward_2c4d14
+               OpReturn
+               OpFunctionEnd
diff --git a/test/tint/builtins/gen/var/faceForward/2c4d14.wgsl.expected.wgsl b/test/tint/builtins/gen/var/faceForward/2c4d14.wgsl.expected.wgsl
new file mode 100644
index 0000000000..994b33e402
--- /dev/null
+++ b/test/tint/builtins/gen/var/faceForward/2c4d14.wgsl.expected.wgsl
@@ -0,0 +1,22 @@
+fn faceForward_2c4d14() {
+  const arg_0 = vec4(1.0);
+  const arg_1 = vec4(1.0);
+  const arg_2 = vec4(1.0);
+  var res = faceForward(arg_0, arg_1, arg_2);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  faceForward_2c4d14();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  faceForward_2c4d14();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  faceForward_2c4d14();
+}
diff --git a/test/tint/builtins/gen/var/faceForward/b42ef3.wgsl b/test/tint/builtins/gen/var/faceForward/b42ef3.wgsl
new file mode 100644
index 0000000000..c069475d20
--- /dev/null
+++ b/test/tint/builtins/gen/var/faceForward/b42ef3.wgsl
@@ -0,0 +1,46 @@
+// 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 faceForward(vec<2, fa>, vec<2, fa>, vec<2, fa>) -> vec<2, fa>
+fn faceForward_b42ef3() {
+  const arg_0 = vec2(1.);
+  const arg_1 = vec2(1.);
+  const arg_2 = vec2(1.);
+  var res = faceForward(arg_0, arg_1, arg_2);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  faceForward_b42ef3();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  faceForward_b42ef3();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  faceForward_b42ef3();
+}
diff --git a/test/tint/builtins/gen/var/faceForward/b42ef3.wgsl.expected.dxc.hlsl b/test/tint/builtins/gen/var/faceForward/b42ef3.wgsl.expected.dxc.hlsl
new file mode 100644
index 0000000000..a91017dedc
--- /dev/null
+++ b/test/tint/builtins/gen/var/faceForward/b42ef3.wgsl.expected.dxc.hlsl
@@ -0,0 +1,30 @@
+void faceForward_b42ef3() {
+  float2 res = (-1.0f).xx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  faceForward_b42ef3();
+  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() {
+  faceForward_b42ef3();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  faceForward_b42ef3();
+  return;
+}
diff --git a/test/tint/builtins/gen/var/faceForward/b42ef3.wgsl.expected.fxc.hlsl b/test/tint/builtins/gen/var/faceForward/b42ef3.wgsl.expected.fxc.hlsl
new file mode 100644
index 0000000000..a91017dedc
--- /dev/null
+++ b/test/tint/builtins/gen/var/faceForward/b42ef3.wgsl.expected.fxc.hlsl
@@ -0,0 +1,30 @@
+void faceForward_b42ef3() {
+  float2 res = (-1.0f).xx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  faceForward_b42ef3();
+  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() {
+  faceForward_b42ef3();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  faceForward_b42ef3();
+  return;
+}
diff --git a/test/tint/builtins/gen/var/faceForward/b42ef3.wgsl.expected.glsl b/test/tint/builtins/gen/var/faceForward/b42ef3.wgsl.expected.glsl
new file mode 100644
index 0000000000..316c6ff1b4
--- /dev/null
+++ b/test/tint/builtins/gen/var/faceForward/b42ef3.wgsl.expected.glsl
@@ -0,0 +1,49 @@
+#version 310 es
+
+void faceForward_b42ef3() {
+  vec2 res = vec2(-1.0f);
+}
+
+vec4 vertex_main() {
+  faceForward_b42ef3();
+  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 faceForward_b42ef3() {
+  vec2 res = vec2(-1.0f);
+}
+
+void fragment_main() {
+  faceForward_b42ef3();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+void faceForward_b42ef3() {
+  vec2 res = vec2(-1.0f);
+}
+
+void compute_main() {
+  faceForward_b42ef3();
+}
+
+layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
+void main() {
+  compute_main();
+  return;
+}
diff --git a/test/tint/builtins/gen/var/faceForward/b42ef3.wgsl.expected.msl b/test/tint/builtins/gen/var/faceForward/b42ef3.wgsl.expected.msl
new file mode 100644
index 0000000000..6a9646f3ad
--- /dev/null
+++ b/test/tint/builtins/gen/var/faceForward/b42ef3.wgsl.expected.msl
@@ -0,0 +1,33 @@
+#include <metal_stdlib>
+
+using namespace metal;
+void faceForward_b42ef3() {
+  float2 res = float2(-1.0f);
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  faceForward_b42ef3();
+  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() {
+  faceForward_b42ef3();
+  return;
+}
+
+kernel void compute_main() {
+  faceForward_b42ef3();
+  return;
+}
+
diff --git a/test/tint/builtins/gen/var/faceForward/b42ef3.wgsl.expected.spvasm b/test/tint/builtins/gen/var/faceForward/b42ef3.wgsl.expected.spvasm
new file mode 100644
index 0000000000..63df79352f
--- /dev/null
+++ b/test/tint/builtins/gen/var/faceForward/b42ef3.wgsl.expected.spvasm
@@ -0,0 +1,67 @@
+; SPIR-V
+; Version: 1.3
+; Generator: Google Tint Compiler; 0
+; Bound: 33
+; 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 %faceForward_b42ef3 "faceForward_b42ef3"
+               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
+    %v2float = OpTypeVector %float 2
+   %float_n1 = OpConstant %float -1
+         %15 = OpConstantComposite %v2float %float_n1 %float_n1
+%_ptr_Function_v2float = OpTypePointer Function %v2float
+         %18 = OpConstantNull %v2float
+         %19 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%faceForward_b42ef3 = OpFunction %void None %9
+         %12 = OpLabel
+        %res = OpVariable %_ptr_Function_v2float Function %18
+               OpStore %res %15
+               OpReturn
+               OpFunctionEnd
+%vertex_main_inner = OpFunction %v4float None %19
+         %21 = OpLabel
+         %22 = OpFunctionCall %void %faceForward_b42ef3
+               OpReturnValue %5
+               OpFunctionEnd
+%vertex_main = OpFunction %void None %9
+         %24 = OpLabel
+         %25 = OpFunctionCall %v4float %vertex_main_inner
+               OpStore %value %25
+               OpStore %vertex_point_size %float_1
+               OpReturn
+               OpFunctionEnd
+%fragment_main = OpFunction %void None %9
+         %28 = OpLabel
+         %29 = OpFunctionCall %void %faceForward_b42ef3
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %31 = OpLabel
+         %32 = OpFunctionCall %void %faceForward_b42ef3
+               OpReturn
+               OpFunctionEnd
diff --git a/test/tint/builtins/gen/var/faceForward/b42ef3.wgsl.expected.wgsl b/test/tint/builtins/gen/var/faceForward/b42ef3.wgsl.expected.wgsl
new file mode 100644
index 0000000000..2b63c9434f
--- /dev/null
+++ b/test/tint/builtins/gen/var/faceForward/b42ef3.wgsl.expected.wgsl
@@ -0,0 +1,22 @@
+fn faceForward_b42ef3() {
+  const arg_0 = vec2(1.0);
+  const arg_1 = vec2(1.0);
+  const arg_2 = vec2(1.0);
+  var res = faceForward(arg_0, arg_1, arg_2);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  faceForward_b42ef3();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  faceForward_b42ef3();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  faceForward_b42ef3();
+}
diff --git a/test/tint/builtins/gen/var/faceForward/fe522b.wgsl b/test/tint/builtins/gen/var/faceForward/fe522b.wgsl
new file mode 100644
index 0000000000..62d967d913
--- /dev/null
+++ b/test/tint/builtins/gen/var/faceForward/fe522b.wgsl
@@ -0,0 +1,46 @@
+// 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 faceForward(vec<3, fa>, vec<3, fa>, vec<3, fa>) -> vec<3, fa>
+fn faceForward_fe522b() {
+  const arg_0 = vec3(1.);
+  const arg_1 = vec3(1.);
+  const arg_2 = vec3(1.);
+  var res = faceForward(arg_0, arg_1, arg_2);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  faceForward_fe522b();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  faceForward_fe522b();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  faceForward_fe522b();
+}
diff --git a/test/tint/builtins/gen/var/faceForward/fe522b.wgsl.expected.dxc.hlsl b/test/tint/builtins/gen/var/faceForward/fe522b.wgsl.expected.dxc.hlsl
new file mode 100644
index 0000000000..706394b401
--- /dev/null
+++ b/test/tint/builtins/gen/var/faceForward/fe522b.wgsl.expected.dxc.hlsl
@@ -0,0 +1,30 @@
+void faceForward_fe522b() {
+  float3 res = (-1.0f).xxx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  faceForward_fe522b();
+  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() {
+  faceForward_fe522b();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  faceForward_fe522b();
+  return;
+}
diff --git a/test/tint/builtins/gen/var/faceForward/fe522b.wgsl.expected.fxc.hlsl b/test/tint/builtins/gen/var/faceForward/fe522b.wgsl.expected.fxc.hlsl
new file mode 100644
index 0000000000..706394b401
--- /dev/null
+++ b/test/tint/builtins/gen/var/faceForward/fe522b.wgsl.expected.fxc.hlsl
@@ -0,0 +1,30 @@
+void faceForward_fe522b() {
+  float3 res = (-1.0f).xxx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  faceForward_fe522b();
+  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() {
+  faceForward_fe522b();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  faceForward_fe522b();
+  return;
+}
diff --git a/test/tint/builtins/gen/var/faceForward/fe522b.wgsl.expected.glsl b/test/tint/builtins/gen/var/faceForward/fe522b.wgsl.expected.glsl
new file mode 100644
index 0000000000..3e26400a49
--- /dev/null
+++ b/test/tint/builtins/gen/var/faceForward/fe522b.wgsl.expected.glsl
@@ -0,0 +1,49 @@
+#version 310 es
+
+void faceForward_fe522b() {
+  vec3 res = vec3(-1.0f);
+}
+
+vec4 vertex_main() {
+  faceForward_fe522b();
+  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 faceForward_fe522b() {
+  vec3 res = vec3(-1.0f);
+}
+
+void fragment_main() {
+  faceForward_fe522b();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+void faceForward_fe522b() {
+  vec3 res = vec3(-1.0f);
+}
+
+void compute_main() {
+  faceForward_fe522b();
+}
+
+layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
+void main() {
+  compute_main();
+  return;
+}
diff --git a/test/tint/builtins/gen/var/faceForward/fe522b.wgsl.expected.msl b/test/tint/builtins/gen/var/faceForward/fe522b.wgsl.expected.msl
new file mode 100644
index 0000000000..4368ad5048
--- /dev/null
+++ b/test/tint/builtins/gen/var/faceForward/fe522b.wgsl.expected.msl
@@ -0,0 +1,33 @@
+#include <metal_stdlib>
+
+using namespace metal;
+void faceForward_fe522b() {
+  float3 res = float3(-1.0f);
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  faceForward_fe522b();
+  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() {
+  faceForward_fe522b();
+  return;
+}
+
+kernel void compute_main() {
+  faceForward_fe522b();
+  return;
+}
+
diff --git a/test/tint/builtins/gen/var/faceForward/fe522b.wgsl.expected.spvasm b/test/tint/builtins/gen/var/faceForward/fe522b.wgsl.expected.spvasm
new file mode 100644
index 0000000000..d3a5b96ea2
--- /dev/null
+++ b/test/tint/builtins/gen/var/faceForward/fe522b.wgsl.expected.spvasm
@@ -0,0 +1,67 @@
+; SPIR-V
+; Version: 1.3
+; Generator: Google Tint Compiler; 0
+; Bound: 33
+; 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 %faceForward_fe522b "faceForward_fe522b"
+               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
+    %v3float = OpTypeVector %float 3
+   %float_n1 = OpConstant %float -1
+         %15 = OpConstantComposite %v3float %float_n1 %float_n1 %float_n1
+%_ptr_Function_v3float = OpTypePointer Function %v3float
+         %18 = OpConstantNull %v3float
+         %19 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%faceForward_fe522b = OpFunction %void None %9
+         %12 = OpLabel
+        %res = OpVariable %_ptr_Function_v3float Function %18
+               OpStore %res %15
+               OpReturn
+               OpFunctionEnd
+%vertex_main_inner = OpFunction %v4float None %19
+         %21 = OpLabel
+         %22 = OpFunctionCall %void %faceForward_fe522b
+               OpReturnValue %5
+               OpFunctionEnd
+%vertex_main = OpFunction %void None %9
+         %24 = OpLabel
+         %25 = OpFunctionCall %v4float %vertex_main_inner
+               OpStore %value %25
+               OpStore %vertex_point_size %float_1
+               OpReturn
+               OpFunctionEnd
+%fragment_main = OpFunction %void None %9
+         %28 = OpLabel
+         %29 = OpFunctionCall %void %faceForward_fe522b
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %31 = OpLabel
+         %32 = OpFunctionCall %void %faceForward_fe522b
+               OpReturn
+               OpFunctionEnd
diff --git a/test/tint/builtins/gen/var/faceForward/fe522b.wgsl.expected.wgsl b/test/tint/builtins/gen/var/faceForward/fe522b.wgsl.expected.wgsl
new file mode 100644
index 0000000000..61f88c42ba
--- /dev/null
+++ b/test/tint/builtins/gen/var/faceForward/fe522b.wgsl.expected.wgsl
@@ -0,0 +1,22 @@
+fn faceForward_fe522b() {
+  const arg_0 = vec3(1.0);
+  const arg_1 = vec3(1.0);
+  const arg_2 = vec3(1.0);
+  var res = faceForward(arg_0, arg_1, arg_2);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  faceForward_fe522b();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  faceForward_fe522b();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  faceForward_fe522b();
+}