Const eval for `normalize` This CL adds const-eval for the `normalize` builtin. Bug: tint:1581 Change-Id: I6d5ba3e0ba507921137ca90c4caefa9daf88f735 Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/111740 Kokoro: Kokoro <noreply+kokoro@google.com> Reviewed-by: Antonio Maiorano <amaiorano@google.com> Commit-Queue: Dan Sinclair <dsinclair@chromium.org>

commit: 8392a82a40b73685a9832d42556efd7c43e4951e [log] [tgz]
author: dan sinclair <dsinclair@chromium.org> Fri Nov 25 04:25:18 2022 +0000
committer: Dawn LUCI CQ <dawn-scoped@luci-project-accounts.iam.gserviceaccount.com> Fri Nov 25 04:25:18 2022 +0000
tree: 779856ee4d88e76fa7f178a8f7fd4212ffdd68f9
parent: 8243aeda757ff29b151a56993c21caca32906676 [diff]
diff --git a/src/tint/intrinsics.def b/src/tint/intrinsics.def
index ba8a581..a148d4d 100644
--- a/src/tint/intrinsics.def
+++ b/src/tint/intrinsics.def

@@ -506,7 +506,7 @@
 fn mix<N: num, T: f32_f16>(vec<N, T>, vec<N, T>, T) -> vec<N, T>
 @const fn modf<T: fa_f32_f16>(@test_value(-1.5) T) -> __modf_result<T>
 @const fn modf<N: num, T: fa_f32_f16>(@test_value(-1.5) vec<N, T>) -> __modf_result_vec<N, T>
-fn normalize<N: num, T: f32_f16>(vec<N, T>) -> vec<N, T>
+@const fn normalize<N: num, T: fa_f32_f16>(vec<N, T>) -> vec<N, T>
 @const fn pack2x16float(vec2<f32>) -> u32
 @const fn pack2x16snorm(vec2<f32>) -> u32
 @const fn pack2x16unorm(vec2<f32>) -> u32

diff --git a/src/tint/resolver/builtin_test.cc b/src/tint/resolver/builtin_test.cc
index ece7656..15bf5de 100644
--- a/src/tint/resolver/builtin_test.cc
+++ b/src/tint/resolver/builtin_test.cc

@@ -1415,7 +1415,7 @@
     EXPECT_EQ(r()->error(), R"(error: no matching call to normalize()
 
 1 candidate function:
-  normalize(vecN<T>) -> vecN<T>  where: T is f32 or f16
+  normalize(vecN<T>) -> vecN<T>  where: T is abstract-float, f32 or f16
 )");
 }
 

diff --git a/src/tint/resolver/const_eval.cc b/src/tint/resolver/const_eval.cc
index 0e6f682..159aaf4 100644
--- a/src/tint/resolver/const_eval.cc
+++ b/src/tint/resolver/const_eval.cc

@@ -461,6 +461,8 @@
 /// CreateElement constructs and returns an Element<T>.
 template <typename T>
 ImplResult CreateElement(ProgramBuilder& builder, const Source& source, const sem::Type* t, T v) {
+    TINT_ASSERT(Resolver, t->is_scalar());
+
     if constexpr (IsFloatingPoint<T>) {
         if (!std::isfinite(v.value)) {
             auto msg = OverflowErrorMessage(v, builder.FriendlyName(t));
@@ -652,8 +654,9 @@
                                    F&& f,
                                    const sem::Constant* c0,
                                    const sem::Constant* c1) {
-    uint32_t n0 = 0, n1 = 0;
+    uint32_t n0 = 0;
     sem::Type::ElementOf(c0->Type(), &n0);
+    uint32_t n1 = 0;
     sem::Type::ElementOf(c1->Type(), &n1);
     uint32_t max_n = std::max(n0, n1);
     // If arity of both constants is 1, invoke callback
@@ -664,7 +667,7 @@
     utils::Vector<const sem::Constant*, 8> els;
     els.Reserve(max_n);
     for (uint32_t i = 0; i < max_n; i++) {
-        auto nested_or_self = [&](auto& c, uint32_t num_elems) {
+        auto nested_or_self = [&](auto* c, uint32_t num_elems) {
             if (num_elems == 1) {
                 return c;
             }
@@ -2734,6 +2737,23 @@
     return CreateComposite(builder, ty, std::move(fields));
 }
 
+ConstEval::Result ConstEval::normalize(const sem::Type* ty,
+                                       utils::VectorRef<const sem::Constant*> args,
+                                       const Source& source) {
+    auto* len_ty = sem::Type::DeepestElementOf(ty);
+    auto len = Length(source, len_ty, args[0]);
+    if (!len) {
+        AddNote("when calculating normalize", source);
+        return utils::Failure;
+    }
+    auto* v = len.Get();
+    if (v->AllZero()) {
+        AddError("zero length vector can not be normalized", source);
+        return utils::Failure;
+    }
+    return OpDivide(ty, utils::Vector{args[0], v}, source);
+}
+
 ConstEval::Result ConstEval::pack2x16float(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 b65bb28..0c8ec6b 100644
--- a/src/tint/resolver/const_eval.h
+++ b/src/tint/resolver/const_eval.h

@@ -719,6 +719,15 @@
                 utils::VectorRef<const sem::Constant*> args,
                 const Source& source);
 
+    /// normalize 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 normalize(const sem::Type* ty,
+                     utils::VectorRef<const sem::Constant*> args,
+                     const Source& source);
+
     /// pack2x16float 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 849f3ef..112d4e2 100644
--- a/src/tint/resolver/const_eval_builtin_test.cc
+++ b/src/tint/resolver/const_eval_builtin_test.cc

@@ -1694,6 +1694,36 @@
                                               ModfCases<f32>(),     //
                                               ModfCases<f16>()))));
 
+template <typename T>
+std::vector<Case> NormalizeCases() {
+    auto error_msg = [&](auto a) {
+        return "12:34 error: " + OverflowErrorMessage(a, "*", a) + R"(
+12:34 note: when calculating normalize)";
+    };
+
+    return {
+        C({Vec(T(2), T(4), T(2))}, Vec(T(0.4082482905), T(0.8164965809), T(0.4082482905)))
+            .FloatComp(),
+
+        C({Vec(T(2), T(0), T(0))}, Vec(T(1), T(0), T(0))),
+        C({Vec(T(0), T(2), T(0))}, Vec(T(0), T(1), T(0))),
+        C({Vec(T(0), T(0), T(2))}, Vec(T(0), T(0), T(1))),
+        C({Vec(-T(2), T(0), T(0))}, Vec(-T(1), T(0), T(0))),
+        C({Vec(T(0), -T(2), T(0))}, Vec(T(0), -T(1), T(0))),
+        C({Vec(T(0), T(0), -T(2))}, Vec(T(0), T(0), -T(1))),
+
+        E({Vec(T(0), T(0), T(0))}, "12:34 error: zero length vector can not be normalized"),
+        E({Vec(T::Highest(), T::Highest(), T::Highest())}, error_msg(T::Highest())),
+    };
+}
+INSTANTIATE_TEST_SUITE_P(  //
+    Normalize,
+    ResolverConstEvalBuiltinTest,
+    testing::Combine(testing::Values(sem::BuiltinType::kNormalize),
+                     testing::ValuesIn(Concat(NormalizeCases<AFloat>(),  //
+                                              NormalizeCases<f32>(),     //
+                                              NormalizeCases<f16>()))));
+
 std::vector<Case> Pack4x8snormCases() {
     return {
         C({Vec(f32(0), f32(0), f32(0), f32(0))}, Val(u32(0x0000'0000))),

diff --git a/src/tint/resolver/intrinsic_table.inl b/src/tint/resolver/intrinsic_table.inl
index 1a2634f..9d77df3 100644
--- a/src/tint/resolver/intrinsic_table.inl
+++ b/src/tint/resolver/intrinsic_table.inl

@@ -13626,12 +13626,12 @@
     /* num parameters */ 1,
     /* num template types */ 1,
     /* num template numbers */ 1,
-    /* template types */ &kTemplateTypes[26],
+    /* template types */ &kTemplateTypes[23],
     /* template numbers */ &kTemplateNumbers[4],
     /* parameters */ &kParameters[880],
     /* return matcher indices */ &kMatcherIndices[30],
     /* flags */ OverloadFlags(OverloadFlag::kIsBuiltin, OverloadFlag::kSupportsVertexPipeline, OverloadFlag::kSupportsFragmentPipeline, OverloadFlag::kSupportsComputePipeline),
-    /* const eval */ nullptr,
+    /* const eval */ &ConstEval::normalize,
   },
   {
     /* [443] */
@@ -14358,7 +14358,7 @@
   },
   {
     /* [54] */
-    /* fn normalize<N : num, T : f32_f16>(vec<N, T>) -> vec<N, T> */
+    /* fn normalize<N : num, T : fa_f32_f16>(vec<N, T>) -> vec<N, T> */
     /* num overloads */ 1,
     /* overloads */ &kOverloads[442],
   },

diff --git a/test/tint/builtins/gen/literal/normalize/39d5ec.wgsl.expected.dxc.hlsl b/test/tint/builtins/gen/literal/normalize/39d5ec.wgsl.expected.dxc.hlsl
index 517a86a..53cdda9 100644
--- a/test/tint/builtins/gen/literal/normalize/39d5ec.wgsl.expected.dxc.hlsl
+++ b/test/tint/builtins/gen/literal/normalize/39d5ec.wgsl.expected.dxc.hlsl

@@ -1,5 +1,5 @@
 void normalize_39d5ec() {
-  vector<float16_t, 3> res = normalize((float16_t(1.0h)).xxx);
+  vector<float16_t, 3> res = (float16_t(0.577148438h)).xxx;
 }
 
 struct tint_symbol {

diff --git a/test/tint/builtins/gen/literal/normalize/39d5ec.wgsl.expected.glsl b/test/tint/builtins/gen/literal/normalize/39d5ec.wgsl.expected.glsl
index 70d93b7..ce3ca5b 100644
--- a/test/tint/builtins/gen/literal/normalize/39d5ec.wgsl.expected.glsl
+++ b/test/tint/builtins/gen/literal/normalize/39d5ec.wgsl.expected.glsl

@@ -2,7 +2,7 @@
 #extension GL_AMD_gpu_shader_half_float : require
 
 void normalize_39d5ec() {
-  f16vec3 res = normalize(f16vec3(1.0hf));
+  f16vec3 res = f16vec3(0.577148438hf);
 }
 
 vec4 vertex_main() {
@@ -23,7 +23,7 @@
 precision mediump float;
 
 void normalize_39d5ec() {
-  f16vec3 res = normalize(f16vec3(1.0hf));
+  f16vec3 res = f16vec3(0.577148438hf);
 }
 
 void fragment_main() {
@@ -38,7 +38,7 @@
 #extension GL_AMD_gpu_shader_half_float : require
 
 void normalize_39d5ec() {
-  f16vec3 res = normalize(f16vec3(1.0hf));
+  f16vec3 res = f16vec3(0.577148438hf);
 }
 
 void compute_main() {

diff --git a/test/tint/builtins/gen/literal/normalize/39d5ec.wgsl.expected.msl b/test/tint/builtins/gen/literal/normalize/39d5ec.wgsl.expected.msl
index 42dca1d..ec7d2f3 100644
--- a/test/tint/builtins/gen/literal/normalize/39d5ec.wgsl.expected.msl
+++ b/test/tint/builtins/gen/literal/normalize/39d5ec.wgsl.expected.msl

@@ -2,7 +2,7 @@
 
 using namespace metal;
 void normalize_39d5ec() {
-  half3 res = normalize(half3(1.0h));
+  half3 res = half3(0.577148438h);
 }
 
 struct tint_symbol {

diff --git a/test/tint/builtins/gen/literal/normalize/39d5ec.wgsl.expected.spvasm b/test/tint/builtins/gen/literal/normalize/39d5ec.wgsl.expected.spvasm
index 0c7ec62..2eefc87 100644
--- a/test/tint/builtins/gen/literal/normalize/39d5ec.wgsl.expected.spvasm
+++ b/test/tint/builtins/gen/literal/normalize/39d5ec.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_0x1_278pn1 = OpConstant %half 0x1.278p-1
+         %16 = OpConstantComposite %v3half %half_0x1_278pn1 %half_0x1_278pn1 %half_0x1_278pn1
 %_ptr_Function_v3half = OpTypePointer Function %v3half
-         %21 = OpConstantNull %v3half
-         %22 = OpTypeFunction %v4float
+         %19 = OpConstantNull %v3half
+         %20 = OpTypeFunction %v4float
     %float_1 = OpConstant %float 1
 %normalize_39d5ec = OpFunction %void None %9
          %12 = OpLabel
-        %res = OpVariable %_ptr_Function_v3half Function %21
-         %13 = OpExtInst %v3half %16 Normalize %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 %normalize_39d5ec
+%vertex_main_inner = OpFunction %v4float None %20
+         %22 = OpLabel
+         %23 = OpFunctionCall %void %normalize_39d5ec
                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 %normalize_39d5ec
+         %29 = OpLabel
+         %30 = OpFunctionCall %void %normalize_39d5ec
                OpReturn
                OpFunctionEnd
 %compute_main = OpFunction %void None %9
-         %34 = OpLabel
-         %35 = OpFunctionCall %void %normalize_39d5ec
+         %32 = OpLabel
+         %33 = OpFunctionCall %void %normalize_39d5ec
                OpReturn
                OpFunctionEnd

diff --git a/test/tint/builtins/gen/literal/normalize/4eaf61.wgsl b/test/tint/builtins/gen/literal/normalize/4eaf61.wgsl
new file mode 100644
index 0000000..fbf5698
--- /dev/null
+++ b/test/tint/builtins/gen/literal/normalize/4eaf61.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 normalize(vec<4, fa>) -> vec<4, fa>
+fn normalize_4eaf61() {
+  var res = normalize(vec4(1.));
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  normalize_4eaf61();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  normalize_4eaf61();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  normalize_4eaf61();
+}

diff --git a/test/tint/builtins/gen/literal/normalize/4eaf61.wgsl.expected.dxc.hlsl b/test/tint/builtins/gen/literal/normalize/4eaf61.wgsl.expected.dxc.hlsl
new file mode 100644
index 0000000..eaa400b
--- /dev/null
+++ b/test/tint/builtins/gen/literal/normalize/4eaf61.wgsl.expected.dxc.hlsl

@@ -0,0 +1,30 @@
+void normalize_4eaf61() {
+  float4 res = (0.5f).xxxx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  normalize_4eaf61();
+  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() {
+  normalize_4eaf61();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  normalize_4eaf61();
+  return;
+}

diff --git a/test/tint/builtins/gen/literal/normalize/4eaf61.wgsl.expected.fxc.hlsl b/test/tint/builtins/gen/literal/normalize/4eaf61.wgsl.expected.fxc.hlsl
new file mode 100644
index 0000000..eaa400b
--- /dev/null
+++ b/test/tint/builtins/gen/literal/normalize/4eaf61.wgsl.expected.fxc.hlsl

@@ -0,0 +1,30 @@
+void normalize_4eaf61() {
+  float4 res = (0.5f).xxxx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  normalize_4eaf61();
+  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() {
+  normalize_4eaf61();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  normalize_4eaf61();
+  return;
+}

diff --git a/test/tint/builtins/gen/literal/normalize/4eaf61.wgsl.expected.glsl b/test/tint/builtins/gen/literal/normalize/4eaf61.wgsl.expected.glsl
new file mode 100644
index 0000000..942c998
--- /dev/null
+++ b/test/tint/builtins/gen/literal/normalize/4eaf61.wgsl.expected.glsl

@@ -0,0 +1,49 @@
+#version 310 es
+
+void normalize_4eaf61() {
+  vec4 res = vec4(0.5f);
+}
+
+vec4 vertex_main() {
+  normalize_4eaf61();
+  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 normalize_4eaf61() {
+  vec4 res = vec4(0.5f);
+}
+
+void fragment_main() {
+  normalize_4eaf61();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+void normalize_4eaf61() {
+  vec4 res = vec4(0.5f);
+}
+
+void compute_main() {
+  normalize_4eaf61();
+}
+
+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/normalize/4eaf61.wgsl.expected.msl b/test/tint/builtins/gen/literal/normalize/4eaf61.wgsl.expected.msl
new file mode 100644
index 0000000..2cd051a
--- /dev/null
+++ b/test/tint/builtins/gen/literal/normalize/4eaf61.wgsl.expected.msl

@@ -0,0 +1,33 @@
+#include <metal_stdlib>
+
+using namespace metal;
+void normalize_4eaf61() {
+  float4 res = float4(0.5f);
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  normalize_4eaf61();
+  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() {
+  normalize_4eaf61();
+  return;
+}
+
+kernel void compute_main() {
+  normalize_4eaf61();
+  return;
+}
+

diff --git a/test/tint/builtins/gen/literal/normalize/4eaf61.wgsl.expected.spvasm b/test/tint/builtins/gen/literal/normalize/4eaf61.wgsl.expected.spvasm
new file mode 100644
index 0000000..83d7867
--- /dev/null
+++ b/test/tint/builtins/gen/literal/normalize/4eaf61.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 %normalize_4eaf61 "normalize_4eaf61"
+               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_0_5 = OpConstant %float 0.5
+         %14 = OpConstantComposite %v4float %float_0_5 %float_0_5 %float_0_5 %float_0_5
+%_ptr_Function_v4float = OpTypePointer Function %v4float
+         %17 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%normalize_4eaf61 = 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 %normalize_4eaf61
+               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 %normalize_4eaf61
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %29 = OpLabel
+         %30 = OpFunctionCall %void %normalize_4eaf61
+               OpReturn
+               OpFunctionEnd

diff --git a/test/tint/builtins/gen/literal/normalize/4eaf61.wgsl.expected.wgsl b/test/tint/builtins/gen/literal/normalize/4eaf61.wgsl.expected.wgsl
new file mode 100644
index 0000000..94c8c08
--- /dev/null
+++ b/test/tint/builtins/gen/literal/normalize/4eaf61.wgsl.expected.wgsl

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

diff --git a/test/tint/builtins/gen/literal/normalize/584e47.wgsl b/test/tint/builtins/gen/literal/normalize/584e47.wgsl
new file mode 100644
index 0000000..9a263eb
--- /dev/null
+++ b/test/tint/builtins/gen/literal/normalize/584e47.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 normalize(vec<2, fa>) -> vec<2, fa>
+fn normalize_584e47() {
+  var res = normalize(vec2(1.));
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  normalize_584e47();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  normalize_584e47();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  normalize_584e47();
+}

diff --git a/test/tint/builtins/gen/literal/normalize/584e47.wgsl.expected.dxc.hlsl b/test/tint/builtins/gen/literal/normalize/584e47.wgsl.expected.dxc.hlsl
new file mode 100644
index 0000000..7ecf609
--- /dev/null
+++ b/test/tint/builtins/gen/literal/normalize/584e47.wgsl.expected.dxc.hlsl

@@ -0,0 +1,30 @@
+void normalize_584e47() {
+  float2 res = (0.707106769f).xx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  normalize_584e47();
+  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() {
+  normalize_584e47();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  normalize_584e47();
+  return;
+}

diff --git a/test/tint/builtins/gen/literal/normalize/584e47.wgsl.expected.fxc.hlsl b/test/tint/builtins/gen/literal/normalize/584e47.wgsl.expected.fxc.hlsl
new file mode 100644
index 0000000..7ecf609
--- /dev/null
+++ b/test/tint/builtins/gen/literal/normalize/584e47.wgsl.expected.fxc.hlsl

@@ -0,0 +1,30 @@
+void normalize_584e47() {
+  float2 res = (0.707106769f).xx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  normalize_584e47();
+  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() {
+  normalize_584e47();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  normalize_584e47();
+  return;
+}

diff --git a/test/tint/builtins/gen/literal/normalize/584e47.wgsl.expected.glsl b/test/tint/builtins/gen/literal/normalize/584e47.wgsl.expected.glsl
new file mode 100644
index 0000000..bc0e2ee
--- /dev/null
+++ b/test/tint/builtins/gen/literal/normalize/584e47.wgsl.expected.glsl

@@ -0,0 +1,49 @@
+#version 310 es
+
+void normalize_584e47() {
+  vec2 res = vec2(0.707106769f);
+}
+
+vec4 vertex_main() {
+  normalize_584e47();
+  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 normalize_584e47() {
+  vec2 res = vec2(0.707106769f);
+}
+
+void fragment_main() {
+  normalize_584e47();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+void normalize_584e47() {
+  vec2 res = vec2(0.707106769f);
+}
+
+void compute_main() {
+  normalize_584e47();
+}
+
+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/normalize/584e47.wgsl.expected.msl b/test/tint/builtins/gen/literal/normalize/584e47.wgsl.expected.msl
new file mode 100644
index 0000000..37095b8
--- /dev/null
+++ b/test/tint/builtins/gen/literal/normalize/584e47.wgsl.expected.msl

@@ -0,0 +1,33 @@
+#include <metal_stdlib>
+
+using namespace metal;
+void normalize_584e47() {
+  float2 res = float2(0.707106769f);
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  normalize_584e47();
+  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() {
+  normalize_584e47();
+  return;
+}
+
+kernel void compute_main() {
+  normalize_584e47();
+  return;
+}
+

diff --git a/test/tint/builtins/gen/literal/normalize/584e47.wgsl.expected.spvasm b/test/tint/builtins/gen/literal/normalize/584e47.wgsl.expected.spvasm
new file mode 100644
index 0000000..3d14b6a
--- /dev/null
+++ b/test/tint/builtins/gen/literal/normalize/584e47.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 %normalize_584e47 "normalize_584e47"
+               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_0_707106769 = OpConstant %float 0.707106769
+         %15 = OpConstantComposite %v2float %float_0_707106769 %float_0_707106769
+%_ptr_Function_v2float = OpTypePointer Function %v2float
+         %18 = OpConstantNull %v2float
+         %19 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%normalize_584e47 = 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 %normalize_584e47
+               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 %normalize_584e47
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %31 = OpLabel
+         %32 = OpFunctionCall %void %normalize_584e47
+               OpReturn
+               OpFunctionEnd

diff --git a/test/tint/builtins/gen/literal/normalize/584e47.wgsl.expected.wgsl b/test/tint/builtins/gen/literal/normalize/584e47.wgsl.expected.wgsl
new file mode 100644
index 0000000..071fecb
--- /dev/null
+++ b/test/tint/builtins/gen/literal/normalize/584e47.wgsl.expected.wgsl

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

diff --git a/test/tint/builtins/gen/literal/normalize/64d8c0.wgsl.expected.dxc.hlsl b/test/tint/builtins/gen/literal/normalize/64d8c0.wgsl.expected.dxc.hlsl
index 435c77f..79592d9 100644
--- a/test/tint/builtins/gen/literal/normalize/64d8c0.wgsl.expected.dxc.hlsl
+++ b/test/tint/builtins/gen/literal/normalize/64d8c0.wgsl.expected.dxc.hlsl

@@ -1,5 +1,5 @@
 void normalize_64d8c0() {
-  float3 res = normalize((1.0f).xxx);
+  float3 res = (0.577350259f).xxx;
 }
 
 struct tint_symbol {

diff --git a/test/tint/builtins/gen/literal/normalize/64d8c0.wgsl.expected.fxc.hlsl b/test/tint/builtins/gen/literal/normalize/64d8c0.wgsl.expected.fxc.hlsl
index 435c77f..79592d9 100644
--- a/test/tint/builtins/gen/literal/normalize/64d8c0.wgsl.expected.fxc.hlsl
+++ b/test/tint/builtins/gen/literal/normalize/64d8c0.wgsl.expected.fxc.hlsl

@@ -1,5 +1,5 @@
 void normalize_64d8c0() {
-  float3 res = normalize((1.0f).xxx);
+  float3 res = (0.577350259f).xxx;
 }
 
 struct tint_symbol {

diff --git a/test/tint/builtins/gen/literal/normalize/64d8c0.wgsl.expected.glsl b/test/tint/builtins/gen/literal/normalize/64d8c0.wgsl.expected.glsl
index ee47fc7..490f3e8 100644
--- a/test/tint/builtins/gen/literal/normalize/64d8c0.wgsl.expected.glsl
+++ b/test/tint/builtins/gen/literal/normalize/64d8c0.wgsl.expected.glsl

@@ -1,7 +1,7 @@
 #version 310 es
 
 void normalize_64d8c0() {
-  vec3 res = normalize(vec3(1.0f));
+  vec3 res = vec3(0.577350259f);
 }
 
 vec4 vertex_main() {
@@ -21,7 +21,7 @@
 precision mediump float;
 
 void normalize_64d8c0() {
-  vec3 res = normalize(vec3(1.0f));
+  vec3 res = vec3(0.577350259f);
 }
 
 void fragment_main() {
@@ -35,7 +35,7 @@
 #version 310 es
 
 void normalize_64d8c0() {
-  vec3 res = normalize(vec3(1.0f));
+  vec3 res = vec3(0.577350259f);
 }
 
 void compute_main() {

diff --git a/test/tint/builtins/gen/literal/normalize/64d8c0.wgsl.expected.msl b/test/tint/builtins/gen/literal/normalize/64d8c0.wgsl.expected.msl
index a821d6d..b7d517c 100644
--- a/test/tint/builtins/gen/literal/normalize/64d8c0.wgsl.expected.msl
+++ b/test/tint/builtins/gen/literal/normalize/64d8c0.wgsl.expected.msl

@@ -2,7 +2,7 @@
 
 using namespace metal;
 void normalize_64d8c0() {
-  float3 res = normalize(float3(1.0f));
+  float3 res = float3(0.577350259f);
 }
 
 struct tint_symbol {

diff --git a/test/tint/builtins/gen/literal/normalize/64d8c0.wgsl.expected.spvasm b/test/tint/builtins/gen/literal/normalize/64d8c0.wgsl.expected.spvasm
index e8caa64..1c825df 100644
--- a/test/tint/builtins/gen/literal/normalize/64d8c0.wgsl.expected.spvasm
+++ b/test/tint/builtins/gen/literal/normalize/64d8c0.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_0_577350259 = OpConstant %float 0.577350259
+         %15 = OpConstantComposite %v3float %float_0_577350259 %float_0_577350259 %float_0_577350259
 %_ptr_Function_v3float = OpTypePointer Function %v3float
-         %20 = OpConstantNull %v3float
-         %21 = OpTypeFunction %v4float
+         %18 = OpConstantNull %v3float
+         %19 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
 %normalize_64d8c0 = OpFunction %void None %9
          %12 = OpLabel
-        %res = OpVariable %_ptr_Function_v3float Function %20
-         %13 = OpExtInst %v3float %15 Normalize %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 %normalize_64d8c0
+%vertex_main_inner = OpFunction %v4float None %19
+         %21 = OpLabel
+         %22 = OpFunctionCall %void %normalize_64d8c0
                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 %normalize_64d8c0
+         %28 = OpLabel
+         %29 = OpFunctionCall %void %normalize_64d8c0
                OpReturn
                OpFunctionEnd
 %compute_main = OpFunction %void None %9
-         %32 = OpLabel
-         %33 = OpFunctionCall %void %normalize_64d8c0
+         %31 = OpLabel
+         %32 = OpFunctionCall %void %normalize_64d8c0
                OpReturn
                OpFunctionEnd

diff --git a/test/tint/builtins/gen/literal/normalize/7990f3.wgsl.expected.dxc.hlsl b/test/tint/builtins/gen/literal/normalize/7990f3.wgsl.expected.dxc.hlsl
index ee11b6f..f031fc9 100644
--- a/test/tint/builtins/gen/literal/normalize/7990f3.wgsl.expected.dxc.hlsl
+++ b/test/tint/builtins/gen/literal/normalize/7990f3.wgsl.expected.dxc.hlsl

@@ -1,5 +1,5 @@
 void normalize_7990f3() {
-  vector<float16_t, 2> res = normalize((float16_t(1.0h)).xx);
+  vector<float16_t, 2> res = (float16_t(0.70703125h)).xx;
 }
 
 struct tint_symbol {

diff --git a/test/tint/builtins/gen/literal/normalize/7990f3.wgsl.expected.glsl b/test/tint/builtins/gen/literal/normalize/7990f3.wgsl.expected.glsl
index 4546245..7e627a3 100644
--- a/test/tint/builtins/gen/literal/normalize/7990f3.wgsl.expected.glsl
+++ b/test/tint/builtins/gen/literal/normalize/7990f3.wgsl.expected.glsl

@@ -2,7 +2,7 @@
 #extension GL_AMD_gpu_shader_half_float : require
 
 void normalize_7990f3() {
-  f16vec2 res = normalize(f16vec2(1.0hf));
+  f16vec2 res = f16vec2(0.70703125hf);
 }
 
 vec4 vertex_main() {
@@ -23,7 +23,7 @@
 precision mediump float;
 
 void normalize_7990f3() {
-  f16vec2 res = normalize(f16vec2(1.0hf));
+  f16vec2 res = f16vec2(0.70703125hf);
 }
 
 void fragment_main() {
@@ -38,7 +38,7 @@
 #extension GL_AMD_gpu_shader_half_float : require
 
 void normalize_7990f3() {
-  f16vec2 res = normalize(f16vec2(1.0hf));
+  f16vec2 res = f16vec2(0.70703125hf);
 }
 
 void compute_main() {

diff --git a/test/tint/builtins/gen/literal/normalize/7990f3.wgsl.expected.msl b/test/tint/builtins/gen/literal/normalize/7990f3.wgsl.expected.msl
index 1a691aa..12287c0 100644
--- a/test/tint/builtins/gen/literal/normalize/7990f3.wgsl.expected.msl
+++ b/test/tint/builtins/gen/literal/normalize/7990f3.wgsl.expected.msl

@@ -2,7 +2,7 @@
 
 using namespace metal;
 void normalize_7990f3() {
-  half2 res = normalize(half2(1.0h));
+  half2 res = half2(0.70703125h);
 }
 
 struct tint_symbol {

diff --git a/test/tint/builtins/gen/literal/normalize/7990f3.wgsl.expected.spvasm b/test/tint/builtins/gen/literal/normalize/7990f3.wgsl.expected.spvasm
index d38b1e8..9a4dbcd 100644
--- a/test/tint/builtins/gen/literal/normalize/7990f3.wgsl.expected.spvasm
+++ b/test/tint/builtins/gen/literal/normalize/7990f3.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_0x1_6apn1 = OpConstant %half 0x1.6ap-1
+         %16 = OpConstantComposite %v2half %half_0x1_6apn1 %half_0x1_6apn1
 %_ptr_Function_v2half = OpTypePointer Function %v2half
-         %21 = OpConstantNull %v2half
-         %22 = OpTypeFunction %v4float
+         %19 = OpConstantNull %v2half
+         %20 = OpTypeFunction %v4float
     %float_1 = OpConstant %float 1
 %normalize_7990f3 = OpFunction %void None %9
          %12 = OpLabel
-        %res = OpVariable %_ptr_Function_v2half Function %21
-         %13 = OpExtInst %v2half %16 Normalize %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 %normalize_7990f3
+%vertex_main_inner = OpFunction %v4float None %20
+         %22 = OpLabel
+         %23 = OpFunctionCall %void %normalize_7990f3
                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 %normalize_7990f3
+         %29 = OpLabel
+         %30 = OpFunctionCall %void %normalize_7990f3
                OpReturn
                OpFunctionEnd
 %compute_main = OpFunction %void None %9
-         %34 = OpLabel
-         %35 = OpFunctionCall %void %normalize_7990f3
+         %32 = OpLabel
+         %33 = OpFunctionCall %void %normalize_7990f3
                OpReturn
                OpFunctionEnd

diff --git a/test/tint/builtins/gen/literal/normalize/9a0aab.wgsl.expected.dxc.hlsl b/test/tint/builtins/gen/literal/normalize/9a0aab.wgsl.expected.dxc.hlsl
index c50afa0..aaa870e 100644
--- a/test/tint/builtins/gen/literal/normalize/9a0aab.wgsl.expected.dxc.hlsl
+++ b/test/tint/builtins/gen/literal/normalize/9a0aab.wgsl.expected.dxc.hlsl

@@ -1,5 +1,5 @@
 void normalize_9a0aab() {
-  float4 res = normalize((1.0f).xxxx);
+  float4 res = (0.5f).xxxx;
 }
 
 struct tint_symbol {

diff --git a/test/tint/builtins/gen/literal/normalize/9a0aab.wgsl.expected.fxc.hlsl b/test/tint/builtins/gen/literal/normalize/9a0aab.wgsl.expected.fxc.hlsl
index c50afa0..aaa870e 100644
--- a/test/tint/builtins/gen/literal/normalize/9a0aab.wgsl.expected.fxc.hlsl
+++ b/test/tint/builtins/gen/literal/normalize/9a0aab.wgsl.expected.fxc.hlsl

@@ -1,5 +1,5 @@
 void normalize_9a0aab() {
-  float4 res = normalize((1.0f).xxxx);
+  float4 res = (0.5f).xxxx;
 }
 
 struct tint_symbol {

diff --git a/test/tint/builtins/gen/literal/normalize/9a0aab.wgsl.expected.glsl b/test/tint/builtins/gen/literal/normalize/9a0aab.wgsl.expected.glsl
index 2e9a75d..9132a7f 100644
--- a/test/tint/builtins/gen/literal/normalize/9a0aab.wgsl.expected.glsl
+++ b/test/tint/builtins/gen/literal/normalize/9a0aab.wgsl.expected.glsl

@@ -1,7 +1,7 @@
 #version 310 es
 
 void normalize_9a0aab() {
-  vec4 res = normalize(vec4(1.0f));
+  vec4 res = vec4(0.5f);
 }
 
 vec4 vertex_main() {
@@ -21,7 +21,7 @@
 precision mediump float;
 
 void normalize_9a0aab() {
-  vec4 res = normalize(vec4(1.0f));
+  vec4 res = vec4(0.5f);
 }
 
 void fragment_main() {
@@ -35,7 +35,7 @@
 #version 310 es
 
 void normalize_9a0aab() {
-  vec4 res = normalize(vec4(1.0f));
+  vec4 res = vec4(0.5f);
 }
 
 void compute_main() {

diff --git a/test/tint/builtins/gen/literal/normalize/9a0aab.wgsl.expected.msl b/test/tint/builtins/gen/literal/normalize/9a0aab.wgsl.expected.msl
index c30ba28..c5ef37d 100644
--- a/test/tint/builtins/gen/literal/normalize/9a0aab.wgsl.expected.msl
+++ b/test/tint/builtins/gen/literal/normalize/9a0aab.wgsl.expected.msl

@@ -2,7 +2,7 @@
 
 using namespace metal;
 void normalize_9a0aab() {
-  float4 res = normalize(float4(1.0f));
+  float4 res = float4(0.5f);
 }
 
 struct tint_symbol {

diff --git a/test/tint/builtins/gen/literal/normalize/9a0aab.wgsl.expected.spvasm b/test/tint/builtins/gen/literal/normalize/9a0aab.wgsl.expected.spvasm
index 65d4b55..7aec576 100644
--- a/test/tint/builtins/gen/literal/normalize/9a0aab.wgsl.expected.spvasm
+++ b/test/tint/builtins/gen/literal/normalize/9a0aab.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_0_5 = OpConstant %float 0.5
+         %14 = OpConstantComposite %v4float %float_0_5 %float_0_5 %float_0_5 %float_0_5
 %_ptr_Function_v4float = OpTypePointer Function %v4float
-         %19 = OpTypeFunction %v4float
+         %17 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
 %normalize_9a0aab = OpFunction %void None %9
          %12 = OpLabel
         %res = OpVariable %_ptr_Function_v4float Function %5
-         %13 = OpExtInst %v4float %14 Normalize %16
-               OpStore %res %13
+               OpStore %res %14
                OpReturn
                OpFunctionEnd
-%vertex_main_inner = OpFunction %v4float None %19
-         %21 = OpLabel
-         %22 = OpFunctionCall %void %normalize_9a0aab
+%vertex_main_inner = OpFunction %v4float None %17
+         %19 = OpLabel
+         %20 = OpFunctionCall %void %normalize_9a0aab
                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 %normalize_9a0aab
+         %26 = OpLabel
+         %27 = OpFunctionCall %void %normalize_9a0aab
                OpReturn
                OpFunctionEnd
 %compute_main = OpFunction %void None %9
-         %30 = OpLabel
-         %31 = OpFunctionCall %void %normalize_9a0aab
+         %29 = OpLabel
+         %30 = OpFunctionCall %void %normalize_9a0aab
                OpReturn
                OpFunctionEnd

diff --git a/test/tint/builtins/gen/literal/normalize/b8cb8d.wgsl.expected.dxc.hlsl b/test/tint/builtins/gen/literal/normalize/b8cb8d.wgsl.expected.dxc.hlsl
index df8561e..2c3c1e3 100644
--- a/test/tint/builtins/gen/literal/normalize/b8cb8d.wgsl.expected.dxc.hlsl
+++ b/test/tint/builtins/gen/literal/normalize/b8cb8d.wgsl.expected.dxc.hlsl

@@ -1,5 +1,5 @@
 void normalize_b8cb8d() {
-  vector<float16_t, 4> res = normalize((float16_t(1.0h)).xxxx);
+  vector<float16_t, 4> res = (float16_t(0.5h)).xxxx;
 }
 
 struct tint_symbol {

diff --git a/test/tint/builtins/gen/literal/normalize/b8cb8d.wgsl.expected.glsl b/test/tint/builtins/gen/literal/normalize/b8cb8d.wgsl.expected.glsl
index 727d29a..9200dcd 100644
--- a/test/tint/builtins/gen/literal/normalize/b8cb8d.wgsl.expected.glsl
+++ b/test/tint/builtins/gen/literal/normalize/b8cb8d.wgsl.expected.glsl

@@ -2,7 +2,7 @@
 #extension GL_AMD_gpu_shader_half_float : require
 
 void normalize_b8cb8d() {
-  f16vec4 res = normalize(f16vec4(1.0hf));
+  f16vec4 res = f16vec4(0.5hf);
 }
 
 vec4 vertex_main() {
@@ -23,7 +23,7 @@
 precision mediump float;
 
 void normalize_b8cb8d() {
-  f16vec4 res = normalize(f16vec4(1.0hf));
+  f16vec4 res = f16vec4(0.5hf);
 }
 
 void fragment_main() {
@@ -38,7 +38,7 @@
 #extension GL_AMD_gpu_shader_half_float : require
 
 void normalize_b8cb8d() {
-  f16vec4 res = normalize(f16vec4(1.0hf));
+  f16vec4 res = f16vec4(0.5hf);
 }
 
 void compute_main() {

diff --git a/test/tint/builtins/gen/literal/normalize/b8cb8d.wgsl.expected.msl b/test/tint/builtins/gen/literal/normalize/b8cb8d.wgsl.expected.msl
index ddeaa45..dac83d3 100644
--- a/test/tint/builtins/gen/literal/normalize/b8cb8d.wgsl.expected.msl
+++ b/test/tint/builtins/gen/literal/normalize/b8cb8d.wgsl.expected.msl

@@ -2,7 +2,7 @@
 
 using namespace metal;
 void normalize_b8cb8d() {
-  half4 res = normalize(half4(1.0h));
+  half4 res = half4(0.5h);
 }
 
 struct tint_symbol {

diff --git a/test/tint/builtins/gen/literal/normalize/b8cb8d.wgsl.expected.spvasm b/test/tint/builtins/gen/literal/normalize/b8cb8d.wgsl.expected.spvasm
index 7b84f9b..e79d052 100644
--- a/test/tint/builtins/gen/literal/normalize/b8cb8d.wgsl.expected.spvasm
+++ b/test/tint/builtins/gen/literal/normalize/b8cb8d.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_0x1pn1 = OpConstant %half 0x1p-1
+         %16 = OpConstantComposite %v4half %half_0x1pn1 %half_0x1pn1 %half_0x1pn1 %half_0x1pn1
 %_ptr_Function_v4half = OpTypePointer Function %v4half
-         %21 = OpConstantNull %v4half
-         %22 = OpTypeFunction %v4float
+         %19 = OpConstantNull %v4half
+         %20 = OpTypeFunction %v4float
     %float_1 = OpConstant %float 1
 %normalize_b8cb8d = OpFunction %void None %9
          %12 = OpLabel
-        %res = OpVariable %_ptr_Function_v4half Function %21
-         %13 = OpExtInst %v4half %16 Normalize %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 %normalize_b8cb8d
+%vertex_main_inner = OpFunction %v4float None %20
+         %22 = OpLabel
+         %23 = OpFunctionCall %void %normalize_b8cb8d
                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 %normalize_b8cb8d
+         %29 = OpLabel
+         %30 = OpFunctionCall %void %normalize_b8cb8d
                OpReturn
                OpFunctionEnd
 %compute_main = OpFunction %void None %9
-         %34 = OpLabel
-         %35 = OpFunctionCall %void %normalize_b8cb8d
+         %32 = OpLabel
+         %33 = OpFunctionCall %void %normalize_b8cb8d
                OpReturn
                OpFunctionEnd

diff --git a/test/tint/builtins/gen/literal/normalize/e7def8.wgsl b/test/tint/builtins/gen/literal/normalize/e7def8.wgsl
new file mode 100644
index 0000000..ba825ee
--- /dev/null
+++ b/test/tint/builtins/gen/literal/normalize/e7def8.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 normalize(vec<3, fa>) -> vec<3, fa>
+fn normalize_e7def8() {
+  var res = normalize(vec3(1.));
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  normalize_e7def8();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  normalize_e7def8();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  normalize_e7def8();
+}

diff --git a/test/tint/builtins/gen/literal/normalize/e7def8.wgsl.expected.dxc.hlsl b/test/tint/builtins/gen/literal/normalize/e7def8.wgsl.expected.dxc.hlsl
new file mode 100644
index 0000000..257e626
--- /dev/null
+++ b/test/tint/builtins/gen/literal/normalize/e7def8.wgsl.expected.dxc.hlsl

@@ -0,0 +1,30 @@
+void normalize_e7def8() {
+  float3 res = (0.577350259f).xxx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  normalize_e7def8();
+  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() {
+  normalize_e7def8();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  normalize_e7def8();
+  return;
+}

diff --git a/test/tint/builtins/gen/literal/normalize/e7def8.wgsl.expected.fxc.hlsl b/test/tint/builtins/gen/literal/normalize/e7def8.wgsl.expected.fxc.hlsl
new file mode 100644
index 0000000..257e626
--- /dev/null
+++ b/test/tint/builtins/gen/literal/normalize/e7def8.wgsl.expected.fxc.hlsl

@@ -0,0 +1,30 @@
+void normalize_e7def8() {
+  float3 res = (0.577350259f).xxx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  normalize_e7def8();
+  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() {
+  normalize_e7def8();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  normalize_e7def8();
+  return;
+}

diff --git a/test/tint/builtins/gen/literal/normalize/e7def8.wgsl.expected.glsl b/test/tint/builtins/gen/literal/normalize/e7def8.wgsl.expected.glsl
new file mode 100644
index 0000000..b29cf51
--- /dev/null
+++ b/test/tint/builtins/gen/literal/normalize/e7def8.wgsl.expected.glsl

@@ -0,0 +1,49 @@
+#version 310 es
+
+void normalize_e7def8() {
+  vec3 res = vec3(0.577350259f);
+}
+
+vec4 vertex_main() {
+  normalize_e7def8();
+  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 normalize_e7def8() {
+  vec3 res = vec3(0.577350259f);
+}
+
+void fragment_main() {
+  normalize_e7def8();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+void normalize_e7def8() {
+  vec3 res = vec3(0.577350259f);
+}
+
+void compute_main() {
+  normalize_e7def8();
+}
+
+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/normalize/e7def8.wgsl.expected.msl b/test/tint/builtins/gen/literal/normalize/e7def8.wgsl.expected.msl
new file mode 100644
index 0000000..b07b880
--- /dev/null
+++ b/test/tint/builtins/gen/literal/normalize/e7def8.wgsl.expected.msl

@@ -0,0 +1,33 @@
+#include <metal_stdlib>
+
+using namespace metal;
+void normalize_e7def8() {
+  float3 res = float3(0.577350259f);
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  normalize_e7def8();
+  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() {
+  normalize_e7def8();
+  return;
+}
+
+kernel void compute_main() {
+  normalize_e7def8();
+  return;
+}
+

diff --git a/test/tint/builtins/gen/literal/normalize/e7def8.wgsl.expected.spvasm b/test/tint/builtins/gen/literal/normalize/e7def8.wgsl.expected.spvasm
new file mode 100644
index 0000000..8b0b7e2
--- /dev/null
+++ b/test/tint/builtins/gen/literal/normalize/e7def8.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 %normalize_e7def8 "normalize_e7def8"
+               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_0_577350259 = OpConstant %float 0.577350259
+         %15 = OpConstantComposite %v3float %float_0_577350259 %float_0_577350259 %float_0_577350259
+%_ptr_Function_v3float = OpTypePointer Function %v3float
+         %18 = OpConstantNull %v3float
+         %19 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%normalize_e7def8 = 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 %normalize_e7def8
+               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 %normalize_e7def8
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %31 = OpLabel
+         %32 = OpFunctionCall %void %normalize_e7def8
+               OpReturn
+               OpFunctionEnd

diff --git a/test/tint/builtins/gen/literal/normalize/e7def8.wgsl.expected.wgsl b/test/tint/builtins/gen/literal/normalize/e7def8.wgsl.expected.wgsl
new file mode 100644
index 0000000..678741a
--- /dev/null
+++ b/test/tint/builtins/gen/literal/normalize/e7def8.wgsl.expected.wgsl

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

diff --git a/test/tint/builtins/gen/literal/normalize/fc2ef1.wgsl.expected.dxc.hlsl b/test/tint/builtins/gen/literal/normalize/fc2ef1.wgsl.expected.dxc.hlsl
index 5eadc79..05e023e 100644
--- a/test/tint/builtins/gen/literal/normalize/fc2ef1.wgsl.expected.dxc.hlsl
+++ b/test/tint/builtins/gen/literal/normalize/fc2ef1.wgsl.expected.dxc.hlsl

@@ -1,5 +1,5 @@
 void normalize_fc2ef1() {
-  float2 res = normalize((1.0f).xx);
+  float2 res = (0.707106769f).xx;
 }
 
 struct tint_symbol {

diff --git a/test/tint/builtins/gen/literal/normalize/fc2ef1.wgsl.expected.fxc.hlsl b/test/tint/builtins/gen/literal/normalize/fc2ef1.wgsl.expected.fxc.hlsl
index 5eadc79..05e023e 100644
--- a/test/tint/builtins/gen/literal/normalize/fc2ef1.wgsl.expected.fxc.hlsl
+++ b/test/tint/builtins/gen/literal/normalize/fc2ef1.wgsl.expected.fxc.hlsl

@@ -1,5 +1,5 @@
 void normalize_fc2ef1() {
-  float2 res = normalize((1.0f).xx);
+  float2 res = (0.707106769f).xx;
 }
 
 struct tint_symbol {

diff --git a/test/tint/builtins/gen/literal/normalize/fc2ef1.wgsl.expected.glsl b/test/tint/builtins/gen/literal/normalize/fc2ef1.wgsl.expected.glsl
index a3ae11e..9cac98d 100644
--- a/test/tint/builtins/gen/literal/normalize/fc2ef1.wgsl.expected.glsl
+++ b/test/tint/builtins/gen/literal/normalize/fc2ef1.wgsl.expected.glsl

@@ -1,7 +1,7 @@
 #version 310 es
 
 void normalize_fc2ef1() {
-  vec2 res = normalize(vec2(1.0f));
+  vec2 res = vec2(0.707106769f);
 }
 
 vec4 vertex_main() {
@@ -21,7 +21,7 @@
 precision mediump float;
 
 void normalize_fc2ef1() {
-  vec2 res = normalize(vec2(1.0f));
+  vec2 res = vec2(0.707106769f);
 }
 
 void fragment_main() {
@@ -35,7 +35,7 @@
 #version 310 es
 
 void normalize_fc2ef1() {
-  vec2 res = normalize(vec2(1.0f));
+  vec2 res = vec2(0.707106769f);
 }
 
 void compute_main() {

diff --git a/test/tint/builtins/gen/literal/normalize/fc2ef1.wgsl.expected.msl b/test/tint/builtins/gen/literal/normalize/fc2ef1.wgsl.expected.msl
index 1c9820b..c2c153d 100644
--- a/test/tint/builtins/gen/literal/normalize/fc2ef1.wgsl.expected.msl
+++ b/test/tint/builtins/gen/literal/normalize/fc2ef1.wgsl.expected.msl

@@ -2,7 +2,7 @@
 
 using namespace metal;
 void normalize_fc2ef1() {
-  float2 res = normalize(float2(1.0f));
+  float2 res = float2(0.707106769f);
 }
 
 struct tint_symbol {

diff --git a/test/tint/builtins/gen/literal/normalize/fc2ef1.wgsl.expected.spvasm b/test/tint/builtins/gen/literal/normalize/fc2ef1.wgsl.expected.spvasm
index 1eb6f88..04dd16a 100644
--- a/test/tint/builtins/gen/literal/normalize/fc2ef1.wgsl.expected.spvasm
+++ b/test/tint/builtins/gen/literal/normalize/fc2ef1.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_0_707106769 = OpConstant %float 0.707106769
+         %15 = OpConstantComposite %v2float %float_0_707106769 %float_0_707106769
 %_ptr_Function_v2float = OpTypePointer Function %v2float
-         %20 = OpConstantNull %v2float
-         %21 = OpTypeFunction %v4float
+         %18 = OpConstantNull %v2float
+         %19 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
 %normalize_fc2ef1 = OpFunction %void None %9
          %12 = OpLabel
-        %res = OpVariable %_ptr_Function_v2float Function %20
-         %13 = OpExtInst %v2float %15 Normalize %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 %normalize_fc2ef1
+%vertex_main_inner = OpFunction %v4float None %19
+         %21 = OpLabel
+         %22 = OpFunctionCall %void %normalize_fc2ef1
                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 %normalize_fc2ef1
+         %28 = OpLabel
+         %29 = OpFunctionCall %void %normalize_fc2ef1
                OpReturn
                OpFunctionEnd
 %compute_main = OpFunction %void None %9
-         %32 = OpLabel
-         %33 = OpFunctionCall %void %normalize_fc2ef1
+         %31 = OpLabel
+         %32 = OpFunctionCall %void %normalize_fc2ef1
                OpReturn
                OpFunctionEnd

diff --git a/test/tint/builtins/gen/var/normalize/4eaf61.wgsl b/test/tint/builtins/gen/var/normalize/4eaf61.wgsl
new file mode 100644
index 0000000..08d23e2
--- /dev/null
+++ b/test/tint/builtins/gen/var/normalize/4eaf61.wgsl

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

diff --git a/test/tint/builtins/gen/var/normalize/4eaf61.wgsl.expected.dxc.hlsl b/test/tint/builtins/gen/var/normalize/4eaf61.wgsl.expected.dxc.hlsl
new file mode 100644
index 0000000..eaa400b
--- /dev/null
+++ b/test/tint/builtins/gen/var/normalize/4eaf61.wgsl.expected.dxc.hlsl

@@ -0,0 +1,30 @@
+void normalize_4eaf61() {
+  float4 res = (0.5f).xxxx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  normalize_4eaf61();
+  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() {
+  normalize_4eaf61();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  normalize_4eaf61();
+  return;
+}

diff --git a/test/tint/builtins/gen/var/normalize/4eaf61.wgsl.expected.fxc.hlsl b/test/tint/builtins/gen/var/normalize/4eaf61.wgsl.expected.fxc.hlsl
new file mode 100644
index 0000000..eaa400b
--- /dev/null
+++ b/test/tint/builtins/gen/var/normalize/4eaf61.wgsl.expected.fxc.hlsl

@@ -0,0 +1,30 @@
+void normalize_4eaf61() {
+  float4 res = (0.5f).xxxx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  normalize_4eaf61();
+  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() {
+  normalize_4eaf61();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  normalize_4eaf61();
+  return;
+}

diff --git a/test/tint/builtins/gen/var/normalize/4eaf61.wgsl.expected.glsl b/test/tint/builtins/gen/var/normalize/4eaf61.wgsl.expected.glsl
new file mode 100644
index 0000000..942c998
--- /dev/null
+++ b/test/tint/builtins/gen/var/normalize/4eaf61.wgsl.expected.glsl

@@ -0,0 +1,49 @@
+#version 310 es
+
+void normalize_4eaf61() {
+  vec4 res = vec4(0.5f);
+}
+
+vec4 vertex_main() {
+  normalize_4eaf61();
+  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 normalize_4eaf61() {
+  vec4 res = vec4(0.5f);
+}
+
+void fragment_main() {
+  normalize_4eaf61();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+void normalize_4eaf61() {
+  vec4 res = vec4(0.5f);
+}
+
+void compute_main() {
+  normalize_4eaf61();
+}
+
+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/normalize/4eaf61.wgsl.expected.msl b/test/tint/builtins/gen/var/normalize/4eaf61.wgsl.expected.msl
new file mode 100644
index 0000000..2cd051a
--- /dev/null
+++ b/test/tint/builtins/gen/var/normalize/4eaf61.wgsl.expected.msl

@@ -0,0 +1,33 @@
+#include <metal_stdlib>
+
+using namespace metal;
+void normalize_4eaf61() {
+  float4 res = float4(0.5f);
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  normalize_4eaf61();
+  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() {
+  normalize_4eaf61();
+  return;
+}
+
+kernel void compute_main() {
+  normalize_4eaf61();
+  return;
+}
+

diff --git a/test/tint/builtins/gen/var/normalize/4eaf61.wgsl.expected.spvasm b/test/tint/builtins/gen/var/normalize/4eaf61.wgsl.expected.spvasm
new file mode 100644
index 0000000..83d7867
--- /dev/null
+++ b/test/tint/builtins/gen/var/normalize/4eaf61.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 %normalize_4eaf61 "normalize_4eaf61"
+               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_0_5 = OpConstant %float 0.5
+         %14 = OpConstantComposite %v4float %float_0_5 %float_0_5 %float_0_5 %float_0_5
+%_ptr_Function_v4float = OpTypePointer Function %v4float
+         %17 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%normalize_4eaf61 = 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 %normalize_4eaf61
+               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 %normalize_4eaf61
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %29 = OpLabel
+         %30 = OpFunctionCall %void %normalize_4eaf61
+               OpReturn
+               OpFunctionEnd

diff --git a/test/tint/builtins/gen/var/normalize/4eaf61.wgsl.expected.wgsl b/test/tint/builtins/gen/var/normalize/4eaf61.wgsl.expected.wgsl
new file mode 100644
index 0000000..e768c8d
--- /dev/null
+++ b/test/tint/builtins/gen/var/normalize/4eaf61.wgsl.expected.wgsl

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

diff --git a/test/tint/builtins/gen/var/normalize/584e47.wgsl b/test/tint/builtins/gen/var/normalize/584e47.wgsl
new file mode 100644
index 0000000..bc1a01c
--- /dev/null
+++ b/test/tint/builtins/gen/var/normalize/584e47.wgsl

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

diff --git a/test/tint/builtins/gen/var/normalize/584e47.wgsl.expected.dxc.hlsl b/test/tint/builtins/gen/var/normalize/584e47.wgsl.expected.dxc.hlsl
new file mode 100644
index 0000000..7ecf609
--- /dev/null
+++ b/test/tint/builtins/gen/var/normalize/584e47.wgsl.expected.dxc.hlsl

@@ -0,0 +1,30 @@
+void normalize_584e47() {
+  float2 res = (0.707106769f).xx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  normalize_584e47();
+  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() {
+  normalize_584e47();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  normalize_584e47();
+  return;
+}

diff --git a/test/tint/builtins/gen/var/normalize/584e47.wgsl.expected.fxc.hlsl b/test/tint/builtins/gen/var/normalize/584e47.wgsl.expected.fxc.hlsl
new file mode 100644
index 0000000..7ecf609
--- /dev/null
+++ b/test/tint/builtins/gen/var/normalize/584e47.wgsl.expected.fxc.hlsl

@@ -0,0 +1,30 @@
+void normalize_584e47() {
+  float2 res = (0.707106769f).xx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  normalize_584e47();
+  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() {
+  normalize_584e47();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  normalize_584e47();
+  return;
+}

diff --git a/test/tint/builtins/gen/var/normalize/584e47.wgsl.expected.glsl b/test/tint/builtins/gen/var/normalize/584e47.wgsl.expected.glsl
new file mode 100644
index 0000000..bc0e2ee
--- /dev/null
+++ b/test/tint/builtins/gen/var/normalize/584e47.wgsl.expected.glsl

@@ -0,0 +1,49 @@
+#version 310 es
+
+void normalize_584e47() {
+  vec2 res = vec2(0.707106769f);
+}
+
+vec4 vertex_main() {
+  normalize_584e47();
+  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 normalize_584e47() {
+  vec2 res = vec2(0.707106769f);
+}
+
+void fragment_main() {
+  normalize_584e47();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+void normalize_584e47() {
+  vec2 res = vec2(0.707106769f);
+}
+
+void compute_main() {
+  normalize_584e47();
+}
+
+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/normalize/584e47.wgsl.expected.msl b/test/tint/builtins/gen/var/normalize/584e47.wgsl.expected.msl
new file mode 100644
index 0000000..37095b8
--- /dev/null
+++ b/test/tint/builtins/gen/var/normalize/584e47.wgsl.expected.msl

@@ -0,0 +1,33 @@
+#include <metal_stdlib>
+
+using namespace metal;
+void normalize_584e47() {
+  float2 res = float2(0.707106769f);
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  normalize_584e47();
+  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() {
+  normalize_584e47();
+  return;
+}
+
+kernel void compute_main() {
+  normalize_584e47();
+  return;
+}
+

diff --git a/test/tint/builtins/gen/var/normalize/584e47.wgsl.expected.spvasm b/test/tint/builtins/gen/var/normalize/584e47.wgsl.expected.spvasm
new file mode 100644
index 0000000..3d14b6a
--- /dev/null
+++ b/test/tint/builtins/gen/var/normalize/584e47.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 %normalize_584e47 "normalize_584e47"
+               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_0_707106769 = OpConstant %float 0.707106769
+         %15 = OpConstantComposite %v2float %float_0_707106769 %float_0_707106769
+%_ptr_Function_v2float = OpTypePointer Function %v2float
+         %18 = OpConstantNull %v2float
+         %19 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%normalize_584e47 = 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 %normalize_584e47
+               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 %normalize_584e47
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %31 = OpLabel
+         %32 = OpFunctionCall %void %normalize_584e47
+               OpReturn
+               OpFunctionEnd

diff --git a/test/tint/builtins/gen/var/normalize/584e47.wgsl.expected.wgsl b/test/tint/builtins/gen/var/normalize/584e47.wgsl.expected.wgsl
new file mode 100644
index 0000000..717d718
--- /dev/null
+++ b/test/tint/builtins/gen/var/normalize/584e47.wgsl.expected.wgsl

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

diff --git a/test/tint/builtins/gen/var/normalize/e7def8.wgsl b/test/tint/builtins/gen/var/normalize/e7def8.wgsl
new file mode 100644
index 0000000..a87e995
--- /dev/null
+++ b/test/tint/builtins/gen/var/normalize/e7def8.wgsl

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

diff --git a/test/tint/builtins/gen/var/normalize/e7def8.wgsl.expected.dxc.hlsl b/test/tint/builtins/gen/var/normalize/e7def8.wgsl.expected.dxc.hlsl
new file mode 100644
index 0000000..257e626
--- /dev/null
+++ b/test/tint/builtins/gen/var/normalize/e7def8.wgsl.expected.dxc.hlsl

@@ -0,0 +1,30 @@
+void normalize_e7def8() {
+  float3 res = (0.577350259f).xxx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  normalize_e7def8();
+  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() {
+  normalize_e7def8();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  normalize_e7def8();
+  return;
+}

diff --git a/test/tint/builtins/gen/var/normalize/e7def8.wgsl.expected.fxc.hlsl b/test/tint/builtins/gen/var/normalize/e7def8.wgsl.expected.fxc.hlsl
new file mode 100644
index 0000000..257e626
--- /dev/null
+++ b/test/tint/builtins/gen/var/normalize/e7def8.wgsl.expected.fxc.hlsl

@@ -0,0 +1,30 @@
+void normalize_e7def8() {
+  float3 res = (0.577350259f).xxx;
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  normalize_e7def8();
+  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() {
+  normalize_e7def8();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  normalize_e7def8();
+  return;
+}

diff --git a/test/tint/builtins/gen/var/normalize/e7def8.wgsl.expected.glsl b/test/tint/builtins/gen/var/normalize/e7def8.wgsl.expected.glsl
new file mode 100644
index 0000000..b29cf51
--- /dev/null
+++ b/test/tint/builtins/gen/var/normalize/e7def8.wgsl.expected.glsl

@@ -0,0 +1,49 @@
+#version 310 es
+
+void normalize_e7def8() {
+  vec3 res = vec3(0.577350259f);
+}
+
+vec4 vertex_main() {
+  normalize_e7def8();
+  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 normalize_e7def8() {
+  vec3 res = vec3(0.577350259f);
+}
+
+void fragment_main() {
+  normalize_e7def8();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+void normalize_e7def8() {
+  vec3 res = vec3(0.577350259f);
+}
+
+void compute_main() {
+  normalize_e7def8();
+}
+
+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/normalize/e7def8.wgsl.expected.msl b/test/tint/builtins/gen/var/normalize/e7def8.wgsl.expected.msl
new file mode 100644
index 0000000..b07b880
--- /dev/null
+++ b/test/tint/builtins/gen/var/normalize/e7def8.wgsl.expected.msl

@@ -0,0 +1,33 @@
+#include <metal_stdlib>
+
+using namespace metal;
+void normalize_e7def8() {
+  float3 res = float3(0.577350259f);
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  normalize_e7def8();
+  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() {
+  normalize_e7def8();
+  return;
+}
+
+kernel void compute_main() {
+  normalize_e7def8();
+  return;
+}
+

diff --git a/test/tint/builtins/gen/var/normalize/e7def8.wgsl.expected.spvasm b/test/tint/builtins/gen/var/normalize/e7def8.wgsl.expected.spvasm
new file mode 100644
index 0000000..8b0b7e2
--- /dev/null
+++ b/test/tint/builtins/gen/var/normalize/e7def8.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 %normalize_e7def8 "normalize_e7def8"
+               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_0_577350259 = OpConstant %float 0.577350259
+         %15 = OpConstantComposite %v3float %float_0_577350259 %float_0_577350259 %float_0_577350259
+%_ptr_Function_v3float = OpTypePointer Function %v3float
+         %18 = OpConstantNull %v3float
+         %19 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%normalize_e7def8 = 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 %normalize_e7def8
+               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 %normalize_e7def8
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %31 = OpLabel
+         %32 = OpFunctionCall %void %normalize_e7def8
+               OpReturn
+               OpFunctionEnd

diff --git a/test/tint/builtins/gen/var/normalize/e7def8.wgsl.expected.wgsl b/test/tint/builtins/gen/var/normalize/e7def8.wgsl.expected.wgsl
new file mode 100644
index 0000000..1974954
--- /dev/null
+++ b/test/tint/builtins/gen/var/normalize/e7def8.wgsl.expected.wgsl

@@ -0,0 +1,20 @@
+fn normalize_e7def8() {
+  const arg_0 = vec3(1.0);
+  var res = normalize(arg_0);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  normalize_e7def8();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  normalize_e7def8();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  normalize_e7def8();
+}
commit	8392a82a40b73685a9832d42556efd7c43e4951e	[log] [tgz]
author	dan sinclair <dsinclair@chromium.org>	Fri Nov 25 04:25:18 2022 +0000
committer	Dawn LUCI CQ <dawn-scoped@luci-project-accounts.iam.gserviceaccount.com>	Fri Nov 25 04:25:18 2022 +0000
tree	779856ee4d88e76fa7f178a8f7fd4212ffdd68f9
parent	8243aeda757ff29b151a56993c21caca32906676 [diff]