tint: Implement const-eval of frexp()

Also add abstract overloads.

Bug: tint:1581
Fixed: tint:1768
Change-Id: Icda465e0cfe960b77823c2135f0cfe8f82ed394f
Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/111441
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Antonio Maiorano <amaiorano@google.com>
Reviewed-by: Dan Sinclair <dsinclair@chromium.org>
Commit-Queue: Ben Clayton <bclayton@google.com>
diff --git a/test/tint/builtins/gen/var/frexp/34bbfb.wgsl b/test/tint/builtins/gen/var/frexp/34bbfb.wgsl
new file mode 100644
index 0000000..5a57803
--- /dev/null
+++ b/test/tint/builtins/gen/var/frexp/34bbfb.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 frexp(vec<4, fa>) -> __frexp_result_vec<4, fa>
+fn frexp_34bbfb() {
+  const arg_0 = vec4(1.);
+  var res = frexp(arg_0);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  frexp_34bbfb();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  frexp_34bbfb();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  frexp_34bbfb();
+}
diff --git a/test/tint/builtins/gen/var/frexp/34bbfb.wgsl.expected.dxc.hlsl b/test/tint/builtins/gen/var/frexp/34bbfb.wgsl.expected.dxc.hlsl
new file mode 100644
index 0000000..f0571de
--- /dev/null
+++ b/test/tint/builtins/gen/var/frexp/34bbfb.wgsl.expected.dxc.hlsl
@@ -0,0 +1,34 @@
+struct frexp_result_vec4 {
+  float4 fract;
+  int4 exp;
+};
+void frexp_34bbfb() {
+  frexp_result_vec4 res = {(0.5f).xxxx, (1).xxxx};
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  frexp_34bbfb();
+  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() {
+  frexp_34bbfb();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  frexp_34bbfb();
+  return;
+}
diff --git a/test/tint/builtins/gen/var/frexp/34bbfb.wgsl.expected.fxc.hlsl b/test/tint/builtins/gen/var/frexp/34bbfb.wgsl.expected.fxc.hlsl
new file mode 100644
index 0000000..f0571de
--- /dev/null
+++ b/test/tint/builtins/gen/var/frexp/34bbfb.wgsl.expected.fxc.hlsl
@@ -0,0 +1,34 @@
+struct frexp_result_vec4 {
+  float4 fract;
+  int4 exp;
+};
+void frexp_34bbfb() {
+  frexp_result_vec4 res = {(0.5f).xxxx, (1).xxxx};
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  frexp_34bbfb();
+  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() {
+  frexp_34bbfb();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  frexp_34bbfb();
+  return;
+}
diff --git a/test/tint/builtins/gen/var/frexp/34bbfb.wgsl.expected.glsl b/test/tint/builtins/gen/var/frexp/34bbfb.wgsl.expected.glsl
new file mode 100644
index 0000000..b05fb00
--- /dev/null
+++ b/test/tint/builtins/gen/var/frexp/34bbfb.wgsl.expected.glsl
@@ -0,0 +1,67 @@
+#version 310 es
+
+struct frexp_result_vec4 {
+  vec4 fract;
+  ivec4 exp;
+};
+
+
+void frexp_34bbfb() {
+  frexp_result_vec4 res = frexp_result_vec4(vec4(0.5f), ivec4(1));
+}
+
+vec4 vertex_main() {
+  frexp_34bbfb();
+  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;
+
+struct frexp_result_vec4 {
+  vec4 fract;
+  ivec4 exp;
+};
+
+
+void frexp_34bbfb() {
+  frexp_result_vec4 res = frexp_result_vec4(vec4(0.5f), ivec4(1));
+}
+
+void fragment_main() {
+  frexp_34bbfb();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+struct frexp_result_vec4 {
+  vec4 fract;
+  ivec4 exp;
+};
+
+
+void frexp_34bbfb() {
+  frexp_result_vec4 res = frexp_result_vec4(vec4(0.5f), ivec4(1));
+}
+
+void compute_main() {
+  frexp_34bbfb();
+}
+
+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/frexp/34bbfb.wgsl.expected.msl b/test/tint/builtins/gen/var/frexp/34bbfb.wgsl.expected.msl
new file mode 100644
index 0000000..1a5b652
--- /dev/null
+++ b/test/tint/builtins/gen/var/frexp/34bbfb.wgsl.expected.msl
@@ -0,0 +1,38 @@
+#include <metal_stdlib>
+
+using namespace metal;
+
+struct frexp_result_vec4 {
+  float4 fract;
+  int4 exp;
+};
+void frexp_34bbfb() {
+  frexp_result_vec4 res = frexp_result_vec4{.fract=float4(0.5f), .exp=int4(1)};
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  frexp_34bbfb();
+  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() {
+  frexp_34bbfb();
+  return;
+}
+
+kernel void compute_main() {
+  frexp_34bbfb();
+  return;
+}
+
diff --git a/test/tint/builtins/gen/var/frexp/34bbfb.wgsl.expected.spvasm b/test/tint/builtins/gen/var/frexp/34bbfb.wgsl.expected.spvasm
new file mode 100644
index 0000000..146a713
--- /dev/null
+++ b/test/tint/builtins/gen/var/frexp/34bbfb.wgsl.expected.spvasm
@@ -0,0 +1,77 @@
+; SPIR-V
+; Version: 1.3
+; Generator: Google Tint Compiler; 0
+; Bound: 38
+; 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 %frexp_34bbfb "frexp_34bbfb"
+               OpName %__frexp_result_vec4 "__frexp_result_vec4"
+               OpMemberName %__frexp_result_vec4 0 "fract"
+               OpMemberName %__frexp_result_vec4 1 "exp"
+               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
+               OpMemberDecorate %__frexp_result_vec4 0 Offset 0
+               OpMemberDecorate %__frexp_result_vec4 1 Offset 16
+      %float = OpTypeFloat 32
+    %v4float = OpTypeVector %float 4
+%_ptr_Output_v4float = OpTypePointer Output %v4float
+          %5 = OpConstantNull %v4float
+      %value = OpVariable %_ptr_Output_v4float Output %5
+%_ptr_Output_float = OpTypePointer Output %float
+          %8 = OpConstantNull %float
+%vertex_point_size = OpVariable %_ptr_Output_float Output %8
+       %void = OpTypeVoid
+          %9 = OpTypeFunction %void
+        %int = OpTypeInt 32 1
+      %v4int = OpTypeVector %int 4
+%__frexp_result_vec4 = OpTypeStruct %v4float %v4int
+  %float_0_5 = OpConstant %float 0.5
+         %17 = OpConstantComposite %v4float %float_0_5 %float_0_5 %float_0_5 %float_0_5
+      %int_1 = OpConstant %int 1
+         %19 = OpConstantComposite %v4int %int_1 %int_1 %int_1 %int_1
+         %20 = OpConstantComposite %__frexp_result_vec4 %17 %19
+%_ptr_Function___frexp_result_vec4 = OpTypePointer Function %__frexp_result_vec4
+         %23 = OpConstantNull %__frexp_result_vec4
+         %24 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%frexp_34bbfb = OpFunction %void None %9
+         %12 = OpLabel
+        %res = OpVariable %_ptr_Function___frexp_result_vec4 Function %23
+               OpStore %res %20
+               OpReturn
+               OpFunctionEnd
+%vertex_main_inner = OpFunction %v4float None %24
+         %26 = OpLabel
+         %27 = OpFunctionCall %void %frexp_34bbfb
+               OpReturnValue %5
+               OpFunctionEnd
+%vertex_main = OpFunction %void None %9
+         %29 = OpLabel
+         %30 = OpFunctionCall %v4float %vertex_main_inner
+               OpStore %value %30
+               OpStore %vertex_point_size %float_1
+               OpReturn
+               OpFunctionEnd
+%fragment_main = OpFunction %void None %9
+         %33 = OpLabel
+         %34 = OpFunctionCall %void %frexp_34bbfb
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %36 = OpLabel
+         %37 = OpFunctionCall %void %frexp_34bbfb
+               OpReturn
+               OpFunctionEnd
diff --git a/test/tint/builtins/gen/var/frexp/34bbfb.wgsl.expected.wgsl b/test/tint/builtins/gen/var/frexp/34bbfb.wgsl.expected.wgsl
new file mode 100644
index 0000000..cb64b2b
--- /dev/null
+++ b/test/tint/builtins/gen/var/frexp/34bbfb.wgsl.expected.wgsl
@@ -0,0 +1,20 @@
+fn frexp_34bbfb() {
+  const arg_0 = vec4(1.0);
+  var res = frexp(arg_0);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  frexp_34bbfb();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  frexp_34bbfb();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  frexp_34bbfb();
+}
diff --git a/test/tint/builtins/gen/var/frexp/6fb3ad.wgsl b/test/tint/builtins/gen/var/frexp/6fb3ad.wgsl
new file mode 100644
index 0000000..a780569
--- /dev/null
+++ b/test/tint/builtins/gen/var/frexp/6fb3ad.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 frexp(vec<2, fa>) -> __frexp_result_vec<2, fa>
+fn frexp_6fb3ad() {
+  const arg_0 = vec2(1.);
+  var res = frexp(arg_0);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  frexp_6fb3ad();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  frexp_6fb3ad();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  frexp_6fb3ad();
+}
diff --git a/test/tint/builtins/gen/var/frexp/6fb3ad.wgsl.expected.dxc.hlsl b/test/tint/builtins/gen/var/frexp/6fb3ad.wgsl.expected.dxc.hlsl
new file mode 100644
index 0000000..cc4ff35
--- /dev/null
+++ b/test/tint/builtins/gen/var/frexp/6fb3ad.wgsl.expected.dxc.hlsl
@@ -0,0 +1,34 @@
+struct frexp_result_vec2 {
+  float2 fract;
+  int2 exp;
+};
+void frexp_6fb3ad() {
+  frexp_result_vec2 res = {(0.5f).xx, (1).xx};
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  frexp_6fb3ad();
+  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() {
+  frexp_6fb3ad();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  frexp_6fb3ad();
+  return;
+}
diff --git a/test/tint/builtins/gen/var/frexp/6fb3ad.wgsl.expected.fxc.hlsl b/test/tint/builtins/gen/var/frexp/6fb3ad.wgsl.expected.fxc.hlsl
new file mode 100644
index 0000000..cc4ff35
--- /dev/null
+++ b/test/tint/builtins/gen/var/frexp/6fb3ad.wgsl.expected.fxc.hlsl
@@ -0,0 +1,34 @@
+struct frexp_result_vec2 {
+  float2 fract;
+  int2 exp;
+};
+void frexp_6fb3ad() {
+  frexp_result_vec2 res = {(0.5f).xx, (1).xx};
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  frexp_6fb3ad();
+  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() {
+  frexp_6fb3ad();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  frexp_6fb3ad();
+  return;
+}
diff --git a/test/tint/builtins/gen/var/frexp/6fb3ad.wgsl.expected.glsl b/test/tint/builtins/gen/var/frexp/6fb3ad.wgsl.expected.glsl
new file mode 100644
index 0000000..e5ecc96
--- /dev/null
+++ b/test/tint/builtins/gen/var/frexp/6fb3ad.wgsl.expected.glsl
@@ -0,0 +1,67 @@
+#version 310 es
+
+struct frexp_result_vec2 {
+  vec2 fract;
+  ivec2 exp;
+};
+
+
+void frexp_6fb3ad() {
+  frexp_result_vec2 res = frexp_result_vec2(vec2(0.5f), ivec2(1));
+}
+
+vec4 vertex_main() {
+  frexp_6fb3ad();
+  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;
+
+struct frexp_result_vec2 {
+  vec2 fract;
+  ivec2 exp;
+};
+
+
+void frexp_6fb3ad() {
+  frexp_result_vec2 res = frexp_result_vec2(vec2(0.5f), ivec2(1));
+}
+
+void fragment_main() {
+  frexp_6fb3ad();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+struct frexp_result_vec2 {
+  vec2 fract;
+  ivec2 exp;
+};
+
+
+void frexp_6fb3ad() {
+  frexp_result_vec2 res = frexp_result_vec2(vec2(0.5f), ivec2(1));
+}
+
+void compute_main() {
+  frexp_6fb3ad();
+}
+
+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/frexp/6fb3ad.wgsl.expected.msl b/test/tint/builtins/gen/var/frexp/6fb3ad.wgsl.expected.msl
new file mode 100644
index 0000000..5356433
--- /dev/null
+++ b/test/tint/builtins/gen/var/frexp/6fb3ad.wgsl.expected.msl
@@ -0,0 +1,38 @@
+#include <metal_stdlib>
+
+using namespace metal;
+
+struct frexp_result_vec2 {
+  float2 fract;
+  int2 exp;
+};
+void frexp_6fb3ad() {
+  frexp_result_vec2 res = frexp_result_vec2{.fract=float2(0.5f), .exp=int2(1)};
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  frexp_6fb3ad();
+  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() {
+  frexp_6fb3ad();
+  return;
+}
+
+kernel void compute_main() {
+  frexp_6fb3ad();
+  return;
+}
+
diff --git a/test/tint/builtins/gen/var/frexp/6fb3ad.wgsl.expected.spvasm b/test/tint/builtins/gen/var/frexp/6fb3ad.wgsl.expected.spvasm
new file mode 100644
index 0000000..1d03915
--- /dev/null
+++ b/test/tint/builtins/gen/var/frexp/6fb3ad.wgsl.expected.spvasm
@@ -0,0 +1,78 @@
+; SPIR-V
+; Version: 1.3
+; Generator: Google Tint Compiler; 0
+; Bound: 39
+; 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 %frexp_6fb3ad "frexp_6fb3ad"
+               OpName %__frexp_result_vec2 "__frexp_result_vec2"
+               OpMemberName %__frexp_result_vec2 0 "fract"
+               OpMemberName %__frexp_result_vec2 1 "exp"
+               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
+               OpMemberDecorate %__frexp_result_vec2 0 Offset 0
+               OpMemberDecorate %__frexp_result_vec2 1 Offset 8
+      %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
+        %int = OpTypeInt 32 1
+      %v2int = OpTypeVector %int 2
+%__frexp_result_vec2 = OpTypeStruct %v2float %v2int
+  %float_0_5 = OpConstant %float 0.5
+         %18 = OpConstantComposite %v2float %float_0_5 %float_0_5
+      %int_1 = OpConstant %int 1
+         %20 = OpConstantComposite %v2int %int_1 %int_1
+         %21 = OpConstantComposite %__frexp_result_vec2 %18 %20
+%_ptr_Function___frexp_result_vec2 = OpTypePointer Function %__frexp_result_vec2
+         %24 = OpConstantNull %__frexp_result_vec2
+         %25 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%frexp_6fb3ad = OpFunction %void None %9
+         %12 = OpLabel
+        %res = OpVariable %_ptr_Function___frexp_result_vec2 Function %24
+               OpStore %res %21
+               OpReturn
+               OpFunctionEnd
+%vertex_main_inner = OpFunction %v4float None %25
+         %27 = OpLabel
+         %28 = OpFunctionCall %void %frexp_6fb3ad
+               OpReturnValue %5
+               OpFunctionEnd
+%vertex_main = OpFunction %void None %9
+         %30 = OpLabel
+         %31 = OpFunctionCall %v4float %vertex_main_inner
+               OpStore %value %31
+               OpStore %vertex_point_size %float_1
+               OpReturn
+               OpFunctionEnd
+%fragment_main = OpFunction %void None %9
+         %34 = OpLabel
+         %35 = OpFunctionCall %void %frexp_6fb3ad
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %37 = OpLabel
+         %38 = OpFunctionCall %void %frexp_6fb3ad
+               OpReturn
+               OpFunctionEnd
diff --git a/test/tint/builtins/gen/var/frexp/6fb3ad.wgsl.expected.wgsl b/test/tint/builtins/gen/var/frexp/6fb3ad.wgsl.expected.wgsl
new file mode 100644
index 0000000..208d085
--- /dev/null
+++ b/test/tint/builtins/gen/var/frexp/6fb3ad.wgsl.expected.wgsl
@@ -0,0 +1,20 @@
+fn frexp_6fb3ad() {
+  const arg_0 = vec2(1.0);
+  var res = frexp(arg_0);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  frexp_6fb3ad();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  frexp_6fb3ad();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  frexp_6fb3ad();
+}
diff --git a/test/tint/builtins/gen/var/frexp/bee870.wgsl b/test/tint/builtins/gen/var/frexp/bee870.wgsl
new file mode 100644
index 0000000..9d01b83
--- /dev/null
+++ b/test/tint/builtins/gen/var/frexp/bee870.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 frexp(fa) -> __frexp_result<fa>
+fn frexp_bee870() {
+  const arg_0 = 1.;
+  var res = frexp(arg_0);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  frexp_bee870();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  frexp_bee870();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  frexp_bee870();
+}
diff --git a/test/tint/builtins/gen/var/frexp/bee870.wgsl.expected.dxc.hlsl b/test/tint/builtins/gen/var/frexp/bee870.wgsl.expected.dxc.hlsl
new file mode 100644
index 0000000..efbe069
--- /dev/null
+++ b/test/tint/builtins/gen/var/frexp/bee870.wgsl.expected.dxc.hlsl
@@ -0,0 +1,34 @@
+struct frexp_result {
+  float fract;
+  int exp;
+};
+void frexp_bee870() {
+  frexp_result res = {0.5f, 1};
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  frexp_bee870();
+  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() {
+  frexp_bee870();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  frexp_bee870();
+  return;
+}
diff --git a/test/tint/builtins/gen/var/frexp/bee870.wgsl.expected.fxc.hlsl b/test/tint/builtins/gen/var/frexp/bee870.wgsl.expected.fxc.hlsl
new file mode 100644
index 0000000..efbe069
--- /dev/null
+++ b/test/tint/builtins/gen/var/frexp/bee870.wgsl.expected.fxc.hlsl
@@ -0,0 +1,34 @@
+struct frexp_result {
+  float fract;
+  int exp;
+};
+void frexp_bee870() {
+  frexp_result res = {0.5f, 1};
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  frexp_bee870();
+  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() {
+  frexp_bee870();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  frexp_bee870();
+  return;
+}
diff --git a/test/tint/builtins/gen/var/frexp/bee870.wgsl.expected.glsl b/test/tint/builtins/gen/var/frexp/bee870.wgsl.expected.glsl
new file mode 100644
index 0000000..0e94ee2
--- /dev/null
+++ b/test/tint/builtins/gen/var/frexp/bee870.wgsl.expected.glsl
@@ -0,0 +1,67 @@
+#version 310 es
+
+struct frexp_result {
+  float fract;
+  int exp;
+};
+
+
+void frexp_bee870() {
+  frexp_result res = frexp_result(0.5f, 1);
+}
+
+vec4 vertex_main() {
+  frexp_bee870();
+  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;
+
+struct frexp_result {
+  float fract;
+  int exp;
+};
+
+
+void frexp_bee870() {
+  frexp_result res = frexp_result(0.5f, 1);
+}
+
+void fragment_main() {
+  frexp_bee870();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+struct frexp_result {
+  float fract;
+  int exp;
+};
+
+
+void frexp_bee870() {
+  frexp_result res = frexp_result(0.5f, 1);
+}
+
+void compute_main() {
+  frexp_bee870();
+}
+
+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/frexp/bee870.wgsl.expected.msl b/test/tint/builtins/gen/var/frexp/bee870.wgsl.expected.msl
new file mode 100644
index 0000000..8c2107d
--- /dev/null
+++ b/test/tint/builtins/gen/var/frexp/bee870.wgsl.expected.msl
@@ -0,0 +1,38 @@
+#include <metal_stdlib>
+
+using namespace metal;
+
+struct frexp_result {
+  float fract;
+  int exp;
+};
+void frexp_bee870() {
+  frexp_result res = frexp_result{.fract=0.5f, .exp=1};
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  frexp_bee870();
+  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() {
+  frexp_bee870();
+  return;
+}
+
+kernel void compute_main() {
+  frexp_bee870();
+  return;
+}
+
diff --git a/test/tint/builtins/gen/var/frexp/bee870.wgsl.expected.spvasm b/test/tint/builtins/gen/var/frexp/bee870.wgsl.expected.spvasm
new file mode 100644
index 0000000..b2675a6
--- /dev/null
+++ b/test/tint/builtins/gen/var/frexp/bee870.wgsl.expected.spvasm
@@ -0,0 +1,74 @@
+; SPIR-V
+; Version: 1.3
+; Generator: Google Tint Compiler; 0
+; Bound: 35
+; 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 %frexp_bee870 "frexp_bee870"
+               OpName %__frexp_result "__frexp_result"
+               OpMemberName %__frexp_result 0 "fract"
+               OpMemberName %__frexp_result 1 "exp"
+               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
+               OpMemberDecorate %__frexp_result 0 Offset 0
+               OpMemberDecorate %__frexp_result 1 Offset 4
+      %float = OpTypeFloat 32
+    %v4float = OpTypeVector %float 4
+%_ptr_Output_v4float = OpTypePointer Output %v4float
+          %5 = OpConstantNull %v4float
+      %value = OpVariable %_ptr_Output_v4float Output %5
+%_ptr_Output_float = OpTypePointer Output %float
+          %8 = OpConstantNull %float
+%vertex_point_size = OpVariable %_ptr_Output_float Output %8
+       %void = OpTypeVoid
+          %9 = OpTypeFunction %void
+        %int = OpTypeInt 32 1
+%__frexp_result = OpTypeStruct %float %int
+  %float_0_5 = OpConstant %float 0.5
+      %int_1 = OpConstant %int 1
+         %17 = OpConstantComposite %__frexp_result %float_0_5 %int_1
+%_ptr_Function___frexp_result = OpTypePointer Function %__frexp_result
+         %20 = OpConstantNull %__frexp_result
+         %21 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%frexp_bee870 = OpFunction %void None %9
+         %12 = OpLabel
+        %res = OpVariable %_ptr_Function___frexp_result Function %20
+               OpStore %res %17
+               OpReturn
+               OpFunctionEnd
+%vertex_main_inner = OpFunction %v4float None %21
+         %23 = OpLabel
+         %24 = OpFunctionCall %void %frexp_bee870
+               OpReturnValue %5
+               OpFunctionEnd
+%vertex_main = OpFunction %void None %9
+         %26 = OpLabel
+         %27 = OpFunctionCall %v4float %vertex_main_inner
+               OpStore %value %27
+               OpStore %vertex_point_size %float_1
+               OpReturn
+               OpFunctionEnd
+%fragment_main = OpFunction %void None %9
+         %30 = OpLabel
+         %31 = OpFunctionCall %void %frexp_bee870
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %33 = OpLabel
+         %34 = OpFunctionCall %void %frexp_bee870
+               OpReturn
+               OpFunctionEnd
diff --git a/test/tint/builtins/gen/var/frexp/bee870.wgsl.expected.wgsl b/test/tint/builtins/gen/var/frexp/bee870.wgsl.expected.wgsl
new file mode 100644
index 0000000..79bda2b
--- /dev/null
+++ b/test/tint/builtins/gen/var/frexp/bee870.wgsl.expected.wgsl
@@ -0,0 +1,20 @@
+fn frexp_bee870() {
+  const arg_0 = 1.0;
+  var res = frexp(arg_0);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  frexp_bee870();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  frexp_bee870();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  frexp_bee870();
+}
diff --git a/test/tint/builtins/gen/var/frexp/bf45ae.wgsl b/test/tint/builtins/gen/var/frexp/bf45ae.wgsl
new file mode 100644
index 0000000..c7477ac
--- /dev/null
+++ b/test/tint/builtins/gen/var/frexp/bf45ae.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 frexp(vec<3, fa>) -> __frexp_result_vec<3, fa>
+fn frexp_bf45ae() {
+  const arg_0 = vec3(1.);
+  var res = frexp(arg_0);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  frexp_bf45ae();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  frexp_bf45ae();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  frexp_bf45ae();
+}
diff --git a/test/tint/builtins/gen/var/frexp/bf45ae.wgsl.expected.dxc.hlsl b/test/tint/builtins/gen/var/frexp/bf45ae.wgsl.expected.dxc.hlsl
new file mode 100644
index 0000000..c896f2b
--- /dev/null
+++ b/test/tint/builtins/gen/var/frexp/bf45ae.wgsl.expected.dxc.hlsl
@@ -0,0 +1,34 @@
+struct frexp_result_vec3 {
+  float3 fract;
+  int3 exp;
+};
+void frexp_bf45ae() {
+  frexp_result_vec3 res = {(0.5f).xxx, (1).xxx};
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  frexp_bf45ae();
+  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() {
+  frexp_bf45ae();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  frexp_bf45ae();
+  return;
+}
diff --git a/test/tint/builtins/gen/var/frexp/bf45ae.wgsl.expected.fxc.hlsl b/test/tint/builtins/gen/var/frexp/bf45ae.wgsl.expected.fxc.hlsl
new file mode 100644
index 0000000..c896f2b
--- /dev/null
+++ b/test/tint/builtins/gen/var/frexp/bf45ae.wgsl.expected.fxc.hlsl
@@ -0,0 +1,34 @@
+struct frexp_result_vec3 {
+  float3 fract;
+  int3 exp;
+};
+void frexp_bf45ae() {
+  frexp_result_vec3 res = {(0.5f).xxx, (1).xxx};
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  frexp_bf45ae();
+  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() {
+  frexp_bf45ae();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  frexp_bf45ae();
+  return;
+}
diff --git a/test/tint/builtins/gen/var/frexp/bf45ae.wgsl.expected.glsl b/test/tint/builtins/gen/var/frexp/bf45ae.wgsl.expected.glsl
new file mode 100644
index 0000000..f7f382e
--- /dev/null
+++ b/test/tint/builtins/gen/var/frexp/bf45ae.wgsl.expected.glsl
@@ -0,0 +1,67 @@
+#version 310 es
+
+struct frexp_result_vec3 {
+  vec3 fract;
+  ivec3 exp;
+};
+
+
+void frexp_bf45ae() {
+  frexp_result_vec3 res = frexp_result_vec3(vec3(0.5f), ivec3(1));
+}
+
+vec4 vertex_main() {
+  frexp_bf45ae();
+  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;
+
+struct frexp_result_vec3 {
+  vec3 fract;
+  ivec3 exp;
+};
+
+
+void frexp_bf45ae() {
+  frexp_result_vec3 res = frexp_result_vec3(vec3(0.5f), ivec3(1));
+}
+
+void fragment_main() {
+  frexp_bf45ae();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+struct frexp_result_vec3 {
+  vec3 fract;
+  ivec3 exp;
+};
+
+
+void frexp_bf45ae() {
+  frexp_result_vec3 res = frexp_result_vec3(vec3(0.5f), ivec3(1));
+}
+
+void compute_main() {
+  frexp_bf45ae();
+}
+
+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/frexp/bf45ae.wgsl.expected.msl b/test/tint/builtins/gen/var/frexp/bf45ae.wgsl.expected.msl
new file mode 100644
index 0000000..6bbab11
--- /dev/null
+++ b/test/tint/builtins/gen/var/frexp/bf45ae.wgsl.expected.msl
@@ -0,0 +1,38 @@
+#include <metal_stdlib>
+
+using namespace metal;
+
+struct frexp_result_vec3 {
+  float3 fract;
+  int3 exp;
+};
+void frexp_bf45ae() {
+  frexp_result_vec3 res = frexp_result_vec3{.fract=float3(0.5f), .exp=int3(1)};
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  frexp_bf45ae();
+  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() {
+  frexp_bf45ae();
+  return;
+}
+
+kernel void compute_main() {
+  frexp_bf45ae();
+  return;
+}
+
diff --git a/test/tint/builtins/gen/var/frexp/bf45ae.wgsl.expected.spvasm b/test/tint/builtins/gen/var/frexp/bf45ae.wgsl.expected.spvasm
new file mode 100644
index 0000000..28459bc
--- /dev/null
+++ b/test/tint/builtins/gen/var/frexp/bf45ae.wgsl.expected.spvasm
@@ -0,0 +1,78 @@
+; SPIR-V
+; Version: 1.3
+; Generator: Google Tint Compiler; 0
+; Bound: 39
+; 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 %frexp_bf45ae "frexp_bf45ae"
+               OpName %__frexp_result_vec3 "__frexp_result_vec3"
+               OpMemberName %__frexp_result_vec3 0 "fract"
+               OpMemberName %__frexp_result_vec3 1 "exp"
+               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
+               OpMemberDecorate %__frexp_result_vec3 0 Offset 0
+               OpMemberDecorate %__frexp_result_vec3 1 Offset 16
+      %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
+        %int = OpTypeInt 32 1
+      %v3int = OpTypeVector %int 3
+%__frexp_result_vec3 = OpTypeStruct %v3float %v3int
+  %float_0_5 = OpConstant %float 0.5
+         %18 = OpConstantComposite %v3float %float_0_5 %float_0_5 %float_0_5
+      %int_1 = OpConstant %int 1
+         %20 = OpConstantComposite %v3int %int_1 %int_1 %int_1
+         %21 = OpConstantComposite %__frexp_result_vec3 %18 %20
+%_ptr_Function___frexp_result_vec3 = OpTypePointer Function %__frexp_result_vec3
+         %24 = OpConstantNull %__frexp_result_vec3
+         %25 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%frexp_bf45ae = OpFunction %void None %9
+         %12 = OpLabel
+        %res = OpVariable %_ptr_Function___frexp_result_vec3 Function %24
+               OpStore %res %21
+               OpReturn
+               OpFunctionEnd
+%vertex_main_inner = OpFunction %v4float None %25
+         %27 = OpLabel
+         %28 = OpFunctionCall %void %frexp_bf45ae
+               OpReturnValue %5
+               OpFunctionEnd
+%vertex_main = OpFunction %void None %9
+         %30 = OpLabel
+         %31 = OpFunctionCall %v4float %vertex_main_inner
+               OpStore %value %31
+               OpStore %vertex_point_size %float_1
+               OpReturn
+               OpFunctionEnd
+%fragment_main = OpFunction %void None %9
+         %34 = OpLabel
+         %35 = OpFunctionCall %void %frexp_bf45ae
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %37 = OpLabel
+         %38 = OpFunctionCall %void %frexp_bf45ae
+               OpReturn
+               OpFunctionEnd
diff --git a/test/tint/builtins/gen/var/frexp/bf45ae.wgsl.expected.wgsl b/test/tint/builtins/gen/var/frexp/bf45ae.wgsl.expected.wgsl
new file mode 100644
index 0000000..0cd674d
--- /dev/null
+++ b/test/tint/builtins/gen/var/frexp/bf45ae.wgsl.expected.wgsl
@@ -0,0 +1,20 @@
+fn frexp_bf45ae() {
+  const arg_0 = vec3(1.0);
+  var res = frexp(arg_0);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  frexp_bf45ae();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  frexp_bf45ae();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  frexp_bf45ae();
+}
diff --git a/test/tint/builtins/gen/var/modf/68d8ee.wgsl.expected.dxc.hlsl b/test/tint/builtins/gen/var/modf/68d8ee.wgsl.expected.dxc.hlsl
new file mode 100644
index 0000000..a09bda3
--- /dev/null
+++ b/test/tint/builtins/gen/var/modf/68d8ee.wgsl.expected.dxc.hlsl
@@ -0,0 +1,34 @@
+struct modf_result_vec3 {
+  float3 fract;
+  float3 whole;
+};
+void modf_68d8ee() {
+  modf_result_vec3 res = {(-0.5f).xxx, (-1.0f).xxx};
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  modf_68d8ee();
+  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() {
+  modf_68d8ee();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  modf_68d8ee();
+  return;
+}
diff --git a/test/tint/builtins/gen/var/modf/68d8ee.wgsl.expected.fxc.hlsl b/test/tint/builtins/gen/var/modf/68d8ee.wgsl.expected.fxc.hlsl
new file mode 100644
index 0000000..a09bda3
--- /dev/null
+++ b/test/tint/builtins/gen/var/modf/68d8ee.wgsl.expected.fxc.hlsl
@@ -0,0 +1,34 @@
+struct modf_result_vec3 {
+  float3 fract;
+  float3 whole;
+};
+void modf_68d8ee() {
+  modf_result_vec3 res = {(-0.5f).xxx, (-1.0f).xxx};
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  modf_68d8ee();
+  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() {
+  modf_68d8ee();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  modf_68d8ee();
+  return;
+}
diff --git a/test/tint/builtins/gen/var/modf/68d8ee.wgsl.expected.glsl b/test/tint/builtins/gen/var/modf/68d8ee.wgsl.expected.glsl
new file mode 100644
index 0000000..24140f3c3
--- /dev/null
+++ b/test/tint/builtins/gen/var/modf/68d8ee.wgsl.expected.glsl
@@ -0,0 +1,67 @@
+#version 310 es
+
+struct modf_result_vec3 {
+  vec3 fract;
+  vec3 whole;
+};
+
+
+void modf_68d8ee() {
+  modf_result_vec3 res = modf_result_vec3(vec3(-0.5f), vec3(-1.0f));
+}
+
+vec4 vertex_main() {
+  modf_68d8ee();
+  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;
+
+struct modf_result_vec3 {
+  vec3 fract;
+  vec3 whole;
+};
+
+
+void modf_68d8ee() {
+  modf_result_vec3 res = modf_result_vec3(vec3(-0.5f), vec3(-1.0f));
+}
+
+void fragment_main() {
+  modf_68d8ee();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+struct modf_result_vec3 {
+  vec3 fract;
+  vec3 whole;
+};
+
+
+void modf_68d8ee() {
+  modf_result_vec3 res = modf_result_vec3(vec3(-0.5f), vec3(-1.0f));
+}
+
+void compute_main() {
+  modf_68d8ee();
+}
+
+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/modf/68d8ee.wgsl.expected.msl b/test/tint/builtins/gen/var/modf/68d8ee.wgsl.expected.msl
new file mode 100644
index 0000000..d24f9fd
--- /dev/null
+++ b/test/tint/builtins/gen/var/modf/68d8ee.wgsl.expected.msl
@@ -0,0 +1,38 @@
+#include <metal_stdlib>
+
+using namespace metal;
+
+struct modf_result_vec3 {
+  float3 fract;
+  float3 whole;
+};
+void modf_68d8ee() {
+  modf_result_vec3 res = modf_result_vec3{.fract=float3(-0.5f), .whole=float3(-1.0f)};
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  modf_68d8ee();
+  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() {
+  modf_68d8ee();
+  return;
+}
+
+kernel void compute_main() {
+  modf_68d8ee();
+  return;
+}
+
diff --git a/test/tint/builtins/gen/var/modf/68d8ee.wgsl.expected.spvasm b/test/tint/builtins/gen/var/modf/68d8ee.wgsl.expected.spvasm
new file mode 100644
index 0000000..e08886d
--- /dev/null
+++ b/test/tint/builtins/gen/var/modf/68d8ee.wgsl.expected.spvasm
@@ -0,0 +1,76 @@
+; SPIR-V
+; Version: 1.3
+; Generator: Google Tint Compiler; 0
+; Bound: 37
+; 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 %modf_68d8ee "modf_68d8ee"
+               OpName %__modf_result_vec3 "__modf_result_vec3"
+               OpMemberName %__modf_result_vec3 0 "fract"
+               OpMemberName %__modf_result_vec3 1 "whole"
+               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
+               OpMemberDecorate %__modf_result_vec3 0 Offset 0
+               OpMemberDecorate %__modf_result_vec3 1 Offset 16
+      %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
+%__modf_result_vec3 = OpTypeStruct %v3float %v3float
+ %float_n0_5 = OpConstant %float -0.5
+         %16 = OpConstantComposite %v3float %float_n0_5 %float_n0_5 %float_n0_5
+   %float_n1 = OpConstant %float -1
+         %18 = OpConstantComposite %v3float %float_n1 %float_n1 %float_n1
+         %19 = OpConstantComposite %__modf_result_vec3 %16 %18
+%_ptr_Function___modf_result_vec3 = OpTypePointer Function %__modf_result_vec3
+         %22 = OpConstantNull %__modf_result_vec3
+         %23 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%modf_68d8ee = OpFunction %void None %9
+         %12 = OpLabel
+        %res = OpVariable %_ptr_Function___modf_result_vec3 Function %22
+               OpStore %res %19
+               OpReturn
+               OpFunctionEnd
+%vertex_main_inner = OpFunction %v4float None %23
+         %25 = OpLabel
+         %26 = OpFunctionCall %void %modf_68d8ee
+               OpReturnValue %5
+               OpFunctionEnd
+%vertex_main = OpFunction %void None %9
+         %28 = OpLabel
+         %29 = OpFunctionCall %v4float %vertex_main_inner
+               OpStore %value %29
+               OpStore %vertex_point_size %float_1
+               OpReturn
+               OpFunctionEnd
+%fragment_main = OpFunction %void None %9
+         %32 = OpLabel
+         %33 = OpFunctionCall %void %modf_68d8ee
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %35 = OpLabel
+         %36 = OpFunctionCall %void %modf_68d8ee
+               OpReturn
+               OpFunctionEnd
diff --git a/test/tint/builtins/gen/var/modf/68d8ee.wgsl.expected.wgsl b/test/tint/builtins/gen/var/modf/68d8ee.wgsl.expected.wgsl
new file mode 100644
index 0000000..c62b099
--- /dev/null
+++ b/test/tint/builtins/gen/var/modf/68d8ee.wgsl.expected.wgsl
@@ -0,0 +1,20 @@
+fn modf_68d8ee() {
+  const arg_0 = vec3(-(1.5));
+  var res = modf(arg_0);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  modf_68d8ee();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  modf_68d8ee();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  modf_68d8ee();
+}
diff --git a/test/tint/builtins/gen/var/modf/732aa6.wgsl.expected.dxc.hlsl b/test/tint/builtins/gen/var/modf/732aa6.wgsl.expected.dxc.hlsl
new file mode 100644
index 0000000..348870d
--- /dev/null
+++ b/test/tint/builtins/gen/var/modf/732aa6.wgsl.expected.dxc.hlsl
@@ -0,0 +1,34 @@
+struct modf_result_vec2 {
+  float2 fract;
+  float2 whole;
+};
+void modf_732aa6() {
+  modf_result_vec2 res = {(-0.5f).xx, (-1.0f).xx};
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  modf_732aa6();
+  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() {
+  modf_732aa6();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  modf_732aa6();
+  return;
+}
diff --git a/test/tint/builtins/gen/var/modf/732aa6.wgsl.expected.fxc.hlsl b/test/tint/builtins/gen/var/modf/732aa6.wgsl.expected.fxc.hlsl
new file mode 100644
index 0000000..348870d
--- /dev/null
+++ b/test/tint/builtins/gen/var/modf/732aa6.wgsl.expected.fxc.hlsl
@@ -0,0 +1,34 @@
+struct modf_result_vec2 {
+  float2 fract;
+  float2 whole;
+};
+void modf_732aa6() {
+  modf_result_vec2 res = {(-0.5f).xx, (-1.0f).xx};
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  modf_732aa6();
+  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() {
+  modf_732aa6();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  modf_732aa6();
+  return;
+}
diff --git a/test/tint/builtins/gen/var/modf/732aa6.wgsl.expected.glsl b/test/tint/builtins/gen/var/modf/732aa6.wgsl.expected.glsl
new file mode 100644
index 0000000..0695a32
--- /dev/null
+++ b/test/tint/builtins/gen/var/modf/732aa6.wgsl.expected.glsl
@@ -0,0 +1,67 @@
+#version 310 es
+
+struct modf_result_vec2 {
+  vec2 fract;
+  vec2 whole;
+};
+
+
+void modf_732aa6() {
+  modf_result_vec2 res = modf_result_vec2(vec2(-0.5f), vec2(-1.0f));
+}
+
+vec4 vertex_main() {
+  modf_732aa6();
+  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;
+
+struct modf_result_vec2 {
+  vec2 fract;
+  vec2 whole;
+};
+
+
+void modf_732aa6() {
+  modf_result_vec2 res = modf_result_vec2(vec2(-0.5f), vec2(-1.0f));
+}
+
+void fragment_main() {
+  modf_732aa6();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+struct modf_result_vec2 {
+  vec2 fract;
+  vec2 whole;
+};
+
+
+void modf_732aa6() {
+  modf_result_vec2 res = modf_result_vec2(vec2(-0.5f), vec2(-1.0f));
+}
+
+void compute_main() {
+  modf_732aa6();
+}
+
+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/modf/732aa6.wgsl.expected.msl b/test/tint/builtins/gen/var/modf/732aa6.wgsl.expected.msl
new file mode 100644
index 0000000..dcb1f40
--- /dev/null
+++ b/test/tint/builtins/gen/var/modf/732aa6.wgsl.expected.msl
@@ -0,0 +1,38 @@
+#include <metal_stdlib>
+
+using namespace metal;
+
+struct modf_result_vec2 {
+  float2 fract;
+  float2 whole;
+};
+void modf_732aa6() {
+  modf_result_vec2 res = modf_result_vec2{.fract=float2(-0.5f), .whole=float2(-1.0f)};
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  modf_732aa6();
+  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() {
+  modf_732aa6();
+  return;
+}
+
+kernel void compute_main() {
+  modf_732aa6();
+  return;
+}
+
diff --git a/test/tint/builtins/gen/var/modf/732aa6.wgsl.expected.spvasm b/test/tint/builtins/gen/var/modf/732aa6.wgsl.expected.spvasm
new file mode 100644
index 0000000..28a7da4
--- /dev/null
+++ b/test/tint/builtins/gen/var/modf/732aa6.wgsl.expected.spvasm
@@ -0,0 +1,76 @@
+; SPIR-V
+; Version: 1.3
+; Generator: Google Tint Compiler; 0
+; Bound: 37
+; 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 %modf_732aa6 "modf_732aa6"
+               OpName %__modf_result_vec2 "__modf_result_vec2"
+               OpMemberName %__modf_result_vec2 0 "fract"
+               OpMemberName %__modf_result_vec2 1 "whole"
+               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
+               OpMemberDecorate %__modf_result_vec2 0 Offset 0
+               OpMemberDecorate %__modf_result_vec2 1 Offset 8
+      %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
+%__modf_result_vec2 = OpTypeStruct %v2float %v2float
+ %float_n0_5 = OpConstant %float -0.5
+         %16 = OpConstantComposite %v2float %float_n0_5 %float_n0_5
+   %float_n1 = OpConstant %float -1
+         %18 = OpConstantComposite %v2float %float_n1 %float_n1
+         %19 = OpConstantComposite %__modf_result_vec2 %16 %18
+%_ptr_Function___modf_result_vec2 = OpTypePointer Function %__modf_result_vec2
+         %22 = OpConstantNull %__modf_result_vec2
+         %23 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%modf_732aa6 = OpFunction %void None %9
+         %12 = OpLabel
+        %res = OpVariable %_ptr_Function___modf_result_vec2 Function %22
+               OpStore %res %19
+               OpReturn
+               OpFunctionEnd
+%vertex_main_inner = OpFunction %v4float None %23
+         %25 = OpLabel
+         %26 = OpFunctionCall %void %modf_732aa6
+               OpReturnValue %5
+               OpFunctionEnd
+%vertex_main = OpFunction %void None %9
+         %28 = OpLabel
+         %29 = OpFunctionCall %v4float %vertex_main_inner
+               OpStore %value %29
+               OpStore %vertex_point_size %float_1
+               OpReturn
+               OpFunctionEnd
+%fragment_main = OpFunction %void None %9
+         %32 = OpLabel
+         %33 = OpFunctionCall %void %modf_732aa6
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %35 = OpLabel
+         %36 = OpFunctionCall %void %modf_732aa6
+               OpReturn
+               OpFunctionEnd
diff --git a/test/tint/builtins/gen/var/modf/732aa6.wgsl.expected.wgsl b/test/tint/builtins/gen/var/modf/732aa6.wgsl.expected.wgsl
new file mode 100644
index 0000000..e1b0b3e
--- /dev/null
+++ b/test/tint/builtins/gen/var/modf/732aa6.wgsl.expected.wgsl
@@ -0,0 +1,20 @@
+fn modf_732aa6() {
+  const arg_0 = vec2(-(1.5));
+  var res = modf(arg_0);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  modf_732aa6();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  modf_732aa6();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  modf_732aa6();
+}
diff --git a/test/tint/builtins/gen/var/modf/c15f48.wgsl.expected.dxc.hlsl b/test/tint/builtins/gen/var/modf/c15f48.wgsl.expected.dxc.hlsl
new file mode 100644
index 0000000..15c5b11
--- /dev/null
+++ b/test/tint/builtins/gen/var/modf/c15f48.wgsl.expected.dxc.hlsl
@@ -0,0 +1,34 @@
+struct modf_result {
+  float fract;
+  float whole;
+};
+void modf_c15f48() {
+  modf_result res = {-0.5f, -1.0f};
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  modf_c15f48();
+  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() {
+  modf_c15f48();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  modf_c15f48();
+  return;
+}
diff --git a/test/tint/builtins/gen/var/modf/c15f48.wgsl.expected.fxc.hlsl b/test/tint/builtins/gen/var/modf/c15f48.wgsl.expected.fxc.hlsl
new file mode 100644
index 0000000..15c5b11
--- /dev/null
+++ b/test/tint/builtins/gen/var/modf/c15f48.wgsl.expected.fxc.hlsl
@@ -0,0 +1,34 @@
+struct modf_result {
+  float fract;
+  float whole;
+};
+void modf_c15f48() {
+  modf_result res = {-0.5f, -1.0f};
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  modf_c15f48();
+  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() {
+  modf_c15f48();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  modf_c15f48();
+  return;
+}
diff --git a/test/tint/builtins/gen/var/modf/c15f48.wgsl.expected.glsl b/test/tint/builtins/gen/var/modf/c15f48.wgsl.expected.glsl
new file mode 100644
index 0000000..3c6410f
--- /dev/null
+++ b/test/tint/builtins/gen/var/modf/c15f48.wgsl.expected.glsl
@@ -0,0 +1,67 @@
+#version 310 es
+
+struct modf_result {
+  float fract;
+  float whole;
+};
+
+
+void modf_c15f48() {
+  modf_result res = modf_result(-0.5f, -1.0f);
+}
+
+vec4 vertex_main() {
+  modf_c15f48();
+  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;
+
+struct modf_result {
+  float fract;
+  float whole;
+};
+
+
+void modf_c15f48() {
+  modf_result res = modf_result(-0.5f, -1.0f);
+}
+
+void fragment_main() {
+  modf_c15f48();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+struct modf_result {
+  float fract;
+  float whole;
+};
+
+
+void modf_c15f48() {
+  modf_result res = modf_result(-0.5f, -1.0f);
+}
+
+void compute_main() {
+  modf_c15f48();
+}
+
+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/modf/c15f48.wgsl.expected.msl b/test/tint/builtins/gen/var/modf/c15f48.wgsl.expected.msl
new file mode 100644
index 0000000..676ddc1
--- /dev/null
+++ b/test/tint/builtins/gen/var/modf/c15f48.wgsl.expected.msl
@@ -0,0 +1,38 @@
+#include <metal_stdlib>
+
+using namespace metal;
+
+struct modf_result {
+  float fract;
+  float whole;
+};
+void modf_c15f48() {
+  modf_result res = modf_result{.fract=-0.5f, .whole=-1.0f};
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  modf_c15f48();
+  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() {
+  modf_c15f48();
+  return;
+}
+
+kernel void compute_main() {
+  modf_c15f48();
+  return;
+}
+
diff --git a/test/tint/builtins/gen/var/modf/c15f48.wgsl.expected.spvasm b/test/tint/builtins/gen/var/modf/c15f48.wgsl.expected.spvasm
new file mode 100644
index 0000000..c18f9c6
--- /dev/null
+++ b/test/tint/builtins/gen/var/modf/c15f48.wgsl.expected.spvasm
@@ -0,0 +1,73 @@
+; SPIR-V
+; Version: 1.3
+; Generator: Google Tint Compiler; 0
+; Bound: 34
+; Schema: 0
+               OpCapability Shader
+               OpMemoryModel Logical GLSL450
+               OpEntryPoint Vertex %vertex_main "vertex_main" %value %vertex_point_size
+               OpEntryPoint Fragment %fragment_main "fragment_main"
+               OpEntryPoint GLCompute %compute_main "compute_main"
+               OpExecutionMode %fragment_main OriginUpperLeft
+               OpExecutionMode %compute_main LocalSize 1 1 1
+               OpName %value "value"
+               OpName %vertex_point_size "vertex_point_size"
+               OpName %modf_c15f48 "modf_c15f48"
+               OpName %__modf_result "__modf_result"
+               OpMemberName %__modf_result 0 "fract"
+               OpMemberName %__modf_result 1 "whole"
+               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
+               OpMemberDecorate %__modf_result 0 Offset 0
+               OpMemberDecorate %__modf_result 1 Offset 4
+      %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
+%__modf_result = OpTypeStruct %float %float
+ %float_n0_5 = OpConstant %float -0.5
+   %float_n1 = OpConstant %float -1
+         %16 = OpConstantComposite %__modf_result %float_n0_5 %float_n1
+%_ptr_Function___modf_result = OpTypePointer Function %__modf_result
+         %19 = OpConstantNull %__modf_result
+         %20 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%modf_c15f48 = OpFunction %void None %9
+         %12 = OpLabel
+        %res = OpVariable %_ptr_Function___modf_result Function %19
+               OpStore %res %16
+               OpReturn
+               OpFunctionEnd
+%vertex_main_inner = OpFunction %v4float None %20
+         %22 = OpLabel
+         %23 = OpFunctionCall %void %modf_c15f48
+               OpReturnValue %5
+               OpFunctionEnd
+%vertex_main = OpFunction %void None %9
+         %25 = OpLabel
+         %26 = OpFunctionCall %v4float %vertex_main_inner
+               OpStore %value %26
+               OpStore %vertex_point_size %float_1
+               OpReturn
+               OpFunctionEnd
+%fragment_main = OpFunction %void None %9
+         %29 = OpLabel
+         %30 = OpFunctionCall %void %modf_c15f48
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %32 = OpLabel
+         %33 = OpFunctionCall %void %modf_c15f48
+               OpReturn
+               OpFunctionEnd
diff --git a/test/tint/builtins/gen/var/modf/c15f48.wgsl.expected.wgsl b/test/tint/builtins/gen/var/modf/c15f48.wgsl.expected.wgsl
new file mode 100644
index 0000000..b343662
--- /dev/null
+++ b/test/tint/builtins/gen/var/modf/c15f48.wgsl.expected.wgsl
@@ -0,0 +1,20 @@
+fn modf_c15f48() {
+  const arg_0 = -(1.5);
+  var res = modf(arg_0);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  modf_c15f48();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  modf_c15f48();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  modf_c15f48();
+}
diff --git a/test/tint/builtins/gen/var/modf/f3d1f9.wgsl.expected.dxc.hlsl b/test/tint/builtins/gen/var/modf/f3d1f9.wgsl.expected.dxc.hlsl
new file mode 100644
index 0000000..e72d03c
--- /dev/null
+++ b/test/tint/builtins/gen/var/modf/f3d1f9.wgsl.expected.dxc.hlsl
@@ -0,0 +1,34 @@
+struct modf_result_vec4 {
+  float4 fract;
+  float4 whole;
+};
+void modf_f3d1f9() {
+  modf_result_vec4 res = {(-0.5f).xxxx, (-1.0f).xxxx};
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  modf_f3d1f9();
+  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() {
+  modf_f3d1f9();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  modf_f3d1f9();
+  return;
+}
diff --git a/test/tint/builtins/gen/var/modf/f3d1f9.wgsl.expected.fxc.hlsl b/test/tint/builtins/gen/var/modf/f3d1f9.wgsl.expected.fxc.hlsl
new file mode 100644
index 0000000..e72d03c
--- /dev/null
+++ b/test/tint/builtins/gen/var/modf/f3d1f9.wgsl.expected.fxc.hlsl
@@ -0,0 +1,34 @@
+struct modf_result_vec4 {
+  float4 fract;
+  float4 whole;
+};
+void modf_f3d1f9() {
+  modf_result_vec4 res = {(-0.5f).xxxx, (-1.0f).xxxx};
+}
+
+struct tint_symbol {
+  float4 value : SV_Position;
+};
+
+float4 vertex_main_inner() {
+  modf_f3d1f9();
+  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() {
+  modf_f3d1f9();
+  return;
+}
+
+[numthreads(1, 1, 1)]
+void compute_main() {
+  modf_f3d1f9();
+  return;
+}
diff --git a/test/tint/builtins/gen/var/modf/f3d1f9.wgsl.expected.glsl b/test/tint/builtins/gen/var/modf/f3d1f9.wgsl.expected.glsl
new file mode 100644
index 0000000..3f2277d
--- /dev/null
+++ b/test/tint/builtins/gen/var/modf/f3d1f9.wgsl.expected.glsl
@@ -0,0 +1,67 @@
+#version 310 es
+
+struct modf_result_vec4 {
+  vec4 fract;
+  vec4 whole;
+};
+
+
+void modf_f3d1f9() {
+  modf_result_vec4 res = modf_result_vec4(vec4(-0.5f), vec4(-1.0f));
+}
+
+vec4 vertex_main() {
+  modf_f3d1f9();
+  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;
+
+struct modf_result_vec4 {
+  vec4 fract;
+  vec4 whole;
+};
+
+
+void modf_f3d1f9() {
+  modf_result_vec4 res = modf_result_vec4(vec4(-0.5f), vec4(-1.0f));
+}
+
+void fragment_main() {
+  modf_f3d1f9();
+}
+
+void main() {
+  fragment_main();
+  return;
+}
+#version 310 es
+
+struct modf_result_vec4 {
+  vec4 fract;
+  vec4 whole;
+};
+
+
+void modf_f3d1f9() {
+  modf_result_vec4 res = modf_result_vec4(vec4(-0.5f), vec4(-1.0f));
+}
+
+void compute_main() {
+  modf_f3d1f9();
+}
+
+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/modf/f3d1f9.wgsl.expected.msl b/test/tint/builtins/gen/var/modf/f3d1f9.wgsl.expected.msl
new file mode 100644
index 0000000..087677a
--- /dev/null
+++ b/test/tint/builtins/gen/var/modf/f3d1f9.wgsl.expected.msl
@@ -0,0 +1,38 @@
+#include <metal_stdlib>
+
+using namespace metal;
+
+struct modf_result_vec4 {
+  float4 fract;
+  float4 whole;
+};
+void modf_f3d1f9() {
+  modf_result_vec4 res = modf_result_vec4{.fract=float4(-0.5f), .whole=float4(-1.0f)};
+}
+
+struct tint_symbol {
+  float4 value [[position]];
+};
+
+float4 vertex_main_inner() {
+  modf_f3d1f9();
+  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() {
+  modf_f3d1f9();
+  return;
+}
+
+kernel void compute_main() {
+  modf_f3d1f9();
+  return;
+}
+
diff --git a/test/tint/builtins/gen/var/modf/f3d1f9.wgsl.expected.spvasm b/test/tint/builtins/gen/var/modf/f3d1f9.wgsl.expected.spvasm
new file mode 100644
index 0000000..629b8a5
--- /dev/null
+++ b/test/tint/builtins/gen/var/modf/f3d1f9.wgsl.expected.spvasm
@@ -0,0 +1,75 @@
+; SPIR-V
+; Version: 1.3
+; Generator: Google Tint Compiler; 0
+; Bound: 36
+; 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 %modf_f3d1f9 "modf_f3d1f9"
+               OpName %__modf_result_vec4 "__modf_result_vec4"
+               OpMemberName %__modf_result_vec4 0 "fract"
+               OpMemberName %__modf_result_vec4 1 "whole"
+               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
+               OpMemberDecorate %__modf_result_vec4 0 Offset 0
+               OpMemberDecorate %__modf_result_vec4 1 Offset 16
+      %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
+%__modf_result_vec4 = OpTypeStruct %v4float %v4float
+ %float_n0_5 = OpConstant %float -0.5
+         %15 = OpConstantComposite %v4float %float_n0_5 %float_n0_5 %float_n0_5 %float_n0_5
+   %float_n1 = OpConstant %float -1
+         %17 = OpConstantComposite %v4float %float_n1 %float_n1 %float_n1 %float_n1
+         %18 = OpConstantComposite %__modf_result_vec4 %15 %17
+%_ptr_Function___modf_result_vec4 = OpTypePointer Function %__modf_result_vec4
+         %21 = OpConstantNull %__modf_result_vec4
+         %22 = OpTypeFunction %v4float
+    %float_1 = OpConstant %float 1
+%modf_f3d1f9 = OpFunction %void None %9
+         %12 = OpLabel
+        %res = OpVariable %_ptr_Function___modf_result_vec4 Function %21
+               OpStore %res %18
+               OpReturn
+               OpFunctionEnd
+%vertex_main_inner = OpFunction %v4float None %22
+         %24 = OpLabel
+         %25 = OpFunctionCall %void %modf_f3d1f9
+               OpReturnValue %5
+               OpFunctionEnd
+%vertex_main = OpFunction %void None %9
+         %27 = OpLabel
+         %28 = OpFunctionCall %v4float %vertex_main_inner
+               OpStore %value %28
+               OpStore %vertex_point_size %float_1
+               OpReturn
+               OpFunctionEnd
+%fragment_main = OpFunction %void None %9
+         %31 = OpLabel
+         %32 = OpFunctionCall %void %modf_f3d1f9
+               OpReturn
+               OpFunctionEnd
+%compute_main = OpFunction %void None %9
+         %34 = OpLabel
+         %35 = OpFunctionCall %void %modf_f3d1f9
+               OpReturn
+               OpFunctionEnd
diff --git a/test/tint/builtins/gen/var/modf/f3d1f9.wgsl.expected.wgsl b/test/tint/builtins/gen/var/modf/f3d1f9.wgsl.expected.wgsl
new file mode 100644
index 0000000..a40cbe8
--- /dev/null
+++ b/test/tint/builtins/gen/var/modf/f3d1f9.wgsl.expected.wgsl
@@ -0,0 +1,20 @@
+fn modf_f3d1f9() {
+  const arg_0 = vec4(-(1.5));
+  var res = modf(arg_0);
+}
+
+@vertex
+fn vertex_main() -> @builtin(position) vec4<f32> {
+  modf_f3d1f9();
+  return vec4<f32>();
+}
+
+@fragment
+fn fragment_main() {
+  modf_f3d1f9();
+}
+
+@compute @workgroup_size(1)
+fn compute_main() {
+  modf_f3d1f9();
+}