writer/msl: Emit builtins as parameters
Add a config parameter for the CanonicalizeEntryPoint transform that
selects between emitting builtins as parameters (for MSL) or struct
members (for HLSL).
This fixes all of the shader IO issues in Tint's E2E tests for MSL.
Fixed: tint:817
Change-Id: Ieb31cdbd2e4d96ac41f8d8515fd07ead8241d770
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/53282
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
Commit-Queue: James Price <jrprice@google.com>
diff --git a/src/transform/canonicalize_entry_point_io.cc b/src/transform/canonicalize_entry_point_io.cc
index 09bd511..d0ac744 100644
--- a/src/transform/canonicalize_entry_point_io.cc
+++ b/src/transform/canonicalize_entry_point_io.cc
@@ -24,6 +24,8 @@
#include "src/sem/struct.h"
#include "src/sem/variable.h"
+TINT_INSTANTIATE_TYPEINFO(tint::transform::CanonicalizeEntryPointIO::Config);
+
namespace tint {
namespace transform {
@@ -60,10 +62,17 @@
} // namespace
-Output CanonicalizeEntryPointIO::Run(const Program* in, const DataMap&) {
+Output CanonicalizeEntryPointIO::Run(const Program* in, const DataMap& data) {
ProgramBuilder out;
CloneContext ctx(&out, in);
+ auto* cfg = data.Get<Config>();
+ if (cfg == nullptr) {
+ out.Diagnostics().add_error(
+ "missing transform data for CanonicalizeEntryPointIO");
+ return Output(Program(std::move(out)));
+ }
+
// Strip entry point IO decorations from struct declarations.
// TODO(jrprice): This code is duplicated with the SPIR-V transform.
for (auto* ty : ctx.src->AST().ConstructedTypes()) {
@@ -104,6 +113,16 @@
auto new_struct_param_symbol = ctx.dst->Sym();
ast::StructMemberList new_struct_members;
for (auto* param : func->Parameters()) {
+ if (cfg->builtin_style == BuiltinStyle::kParameter &&
+ ast::HasDecoration<ast::BuiltinDecoration>(
+ param->Declaration()->decorations())) {
+ // If this parameter is a builtin and we are emitting those as
+ // parameters, then just clone it as is.
+ new_parameters.push_back(
+ ctx.Clone(const_cast<ast::Variable*>(param->Declaration())));
+ continue;
+ }
+
auto param_name = ctx.Clone(param->Declaration()->symbol());
auto* param_ty = param->Type();
auto* param_declared_ty = param->Declaration()->type();
@@ -118,6 +137,20 @@
TINT_ICE(ctx.dst->Diagnostics()) << "nested pipeline IO struct";
}
+ if (cfg->builtin_style == BuiltinStyle::kParameter &&
+ ast::HasDecoration<ast::BuiltinDecoration>(
+ member->Declaration()->decorations())) {
+ // If this struct member is a builtin and we are emitting those as
+ // parameters, then move it to the parameter list.
+ auto* member_ty = CreateASTTypeFor(&ctx, member->Type());
+ auto new_param_name = ctx.dst->Sym();
+ new_parameters.push_back(ctx.dst->Param(
+ new_param_name, member_ty,
+ ctx.Clone(member->Declaration()->decorations())));
+ init_values.push_back(ctx.dst->Expr(new_param_name));
+ continue;
+ }
+
ast::DecorationList new_decorations = RemoveDecorations(
&ctx, member->Declaration()->decorations(),
[](const ast::Decoration* deco) {
@@ -161,21 +194,23 @@
}
}
- // Sort struct members to satisfy HLSL interfacing matching rules.
- std::sort(new_struct_members.begin(), new_struct_members.end(),
- StructMemberComparator);
+ if (!new_struct_members.empty()) {
+ // Sort struct members to satisfy HLSL interfacing matching rules.
+ std::sort(new_struct_members.begin(), new_struct_members.end(),
+ StructMemberComparator);
- // Create the new struct type.
- auto in_struct_name = ctx.dst->Sym();
- auto* in_struct = ctx.dst->create<ast::Struct>(
- in_struct_name, new_struct_members, ast::DecorationList{});
- ctx.InsertBefore(ctx.src->AST().GlobalDeclarations(), func_ast,
- in_struct);
+ // Create the new struct type.
+ auto in_struct_name = ctx.dst->Sym();
+ auto* in_struct = ctx.dst->create<ast::Struct>(
+ in_struct_name, new_struct_members, ast::DecorationList{});
+ ctx.InsertBefore(ctx.src->AST().GlobalDeclarations(), func_ast,
+ in_struct);
- // Create a new function parameter using this struct type.
- auto* struct_param = ctx.dst->Param(
- new_struct_param_symbol, ctx.dst->ty.type_name(in_struct_name));
- new_parameters.push_back(struct_param);
+ // Create a new function parameter using this struct type.
+ auto* struct_param = ctx.dst->Param(
+ new_struct_param_symbol, ctx.dst->ty.type_name(in_struct_name));
+ new_parameters.push_back(struct_param);
+ }
}
// Handle return type.
@@ -273,5 +308,13 @@
return Output(Program(std::move(out)));
}
+CanonicalizeEntryPointIO::Config::Config(BuiltinStyle builtins)
+ : builtin_style(builtins) {}
+
+CanonicalizeEntryPointIO::Config::Config(const Config&) = default;
+CanonicalizeEntryPointIO::Config::~Config() = default;
+CanonicalizeEntryPointIO::Config& CanonicalizeEntryPointIO::Config::operator=(
+ const Config&) = default;
+
} // namespace transform
} // namespace tint
diff --git a/src/transform/canonicalize_entry_point_io.h b/src/transform/canonicalize_entry_point_io.h
index 4298459..3cde217 100644
--- a/src/transform/canonicalize_entry_point_io.h
+++ b/src/transform/canonicalize_entry_point_io.h
@@ -70,6 +70,34 @@
/// ```
class CanonicalizeEntryPointIO : public Transform {
public:
+ /// BuiltinStyle is an enumerator of different ways to emit builtins.
+ enum class BuiltinStyle {
+ /// Use non-struct function parameters for all builtins.
+ kParameter,
+ /// Use struct members for all builtins.
+ kStructMember,
+ };
+
+ /// Configuration options for the transform.
+ struct Config : public Castable<Config, Data> {
+ /// Constructor
+ /// @param builtins the approach to use for emitting builtins.
+ explicit Config(BuiltinStyle builtins);
+
+ /// Copy constructor
+ Config(const Config&);
+
+ /// Destructor
+ ~Config() override;
+
+ /// Assignment operator
+ /// @returns this Config
+ Config& operator=(const Config&);
+
+ /// The approach to use for emitting builtins.
+ BuiltinStyle builtin_style;
+ };
+
/// Constructor
CanonicalizeEntryPointIO();
~CanonicalizeEntryPointIO() override;
diff --git a/src/transform/canonicalize_entry_point_io_test.cc b/src/transform/canonicalize_entry_point_io_test.cc
index e1f32c7..9d2c50d 100644
--- a/src/transform/canonicalize_entry_point_io_test.cc
+++ b/src/transform/canonicalize_entry_point_io_test.cc
@@ -22,7 +22,51 @@
using CanonicalizeEntryPointIOTest = TransformTest;
-TEST_F(CanonicalizeEntryPointIOTest, Parameters) {
+TEST_F(CanonicalizeEntryPointIOTest, Error_MissingTransformData) {
+ auto* src = "";
+
+ auto* expect = "error: missing transform data for CanonicalizeEntryPointIO";
+
+ auto got = Run<CanonicalizeEntryPointIO>(src);
+
+ EXPECT_EQ(expect, str(got));
+}
+
+TEST_F(CanonicalizeEntryPointIOTest, Parameters_BuiltinsAsParameters) {
+ auto* src = R"(
+[[stage(fragment)]]
+fn frag_main([[location(1)]] loc1 : f32,
+ [[location(2)]] loc2 : vec4<u32>,
+ [[builtin(position)]] coord : vec4<f32>) {
+ var col : f32 = (coord.x * loc1);
+}
+)";
+
+ auto* expect = R"(
+struct tint_symbol_1 {
+ [[location(1)]]
+ loc1 : f32;
+ [[location(2)]]
+ loc2 : vec4<u32>;
+};
+
+[[stage(fragment)]]
+fn frag_main([[builtin(position)]] coord : vec4<f32>, tint_symbol : tint_symbol_1) {
+ let loc1 : f32 = tint_symbol.loc1;
+ let loc2 : vec4<u32> = tint_symbol.loc2;
+ var col : f32 = (coord.x * loc1);
+}
+)";
+
+ DataMap data;
+ data.Add<CanonicalizeEntryPointIO::Config>(
+ CanonicalizeEntryPointIO::BuiltinStyle::kParameter);
+ auto got = Run<CanonicalizeEntryPointIO>(src, data);
+
+ EXPECT_EQ(expect, str(got));
+}
+
+TEST_F(CanonicalizeEntryPointIOTest, Parameters_BuiltinsAsStructMembers) {
auto* src = R"(
[[stage(fragment)]]
fn frag_main([[location(1)]] loc1 : f32,
@@ -51,7 +95,10 @@
}
)";
- auto got = Run<CanonicalizeEntryPointIO>(src);
+ DataMap data;
+ data.Add<CanonicalizeEntryPointIO::Config>(
+ CanonicalizeEntryPointIO::BuiltinStyle::kStructMember);
+ auto got = Run<CanonicalizeEntryPointIO>(src, data);
EXPECT_EQ(expect, str(got));
}
@@ -81,12 +128,49 @@
}
)";
- auto got = Run<CanonicalizeEntryPointIO>(src);
+ DataMap data;
+ data.Add<CanonicalizeEntryPointIO::Config>(
+ CanonicalizeEntryPointIO::BuiltinStyle::kParameter);
+ auto got = Run<CanonicalizeEntryPointIO>(src, data);
EXPECT_EQ(expect, str(got));
}
-TEST_F(CanonicalizeEntryPointIOTest, Parameters_EmptyBody) {
+TEST_F(CanonicalizeEntryPointIOTest,
+ Parameters_EmptyBody_BuiltinsAsParameters) {
+ auto* src = R"(
+[[stage(fragment)]]
+fn frag_main([[location(1)]] loc1 : f32,
+ [[location(2)]] loc2 : vec4<u32>,
+ [[builtin(position)]] coord : vec4<f32>) {
+}
+)";
+
+ auto* expect = R"(
+struct tint_symbol_1 {
+ [[location(1)]]
+ loc1 : f32;
+ [[location(2)]]
+ loc2 : vec4<u32>;
+};
+
+[[stage(fragment)]]
+fn frag_main([[builtin(position)]] coord : vec4<f32>, tint_symbol : tint_symbol_1) {
+ let loc1 : f32 = tint_symbol.loc1;
+ let loc2 : vec4<u32> = tint_symbol.loc2;
+}
+)";
+
+ DataMap data;
+ data.Add<CanonicalizeEntryPointIO::Config>(
+ CanonicalizeEntryPointIO::BuiltinStyle::kParameter);
+ auto got = Run<CanonicalizeEntryPointIO>(src, data);
+
+ EXPECT_EQ(expect, str(got));
+}
+
+TEST_F(CanonicalizeEntryPointIOTest,
+ Parameters_EmptyBody_BuiltinsAsStructMembers) {
auto* src = R"(
[[stage(fragment)]]
fn frag_main([[location(1)]] loc1 : f32,
@@ -113,12 +197,69 @@
}
)";
- auto got = Run<CanonicalizeEntryPointIO>(src);
+ DataMap data;
+ data.Add<CanonicalizeEntryPointIO::Config>(
+ CanonicalizeEntryPointIO::BuiltinStyle::kStructMember);
+ auto got = Run<CanonicalizeEntryPointIO>(src, data);
EXPECT_EQ(expect, str(got));
}
-TEST_F(CanonicalizeEntryPointIOTest, StructParameters) {
+TEST_F(CanonicalizeEntryPointIOTest, StructParameters_BuiltinsAsParameters) {
+ auto* src = R"(
+struct FragBuiltins {
+ [[builtin(position)]] coord : vec4<f32>;
+};
+struct FragLocations {
+ [[location(1)]] loc1 : f32;
+ [[location(2)]] loc2 : vec4<u32>;
+};
+
+[[stage(fragment)]]
+fn frag_main([[location(0)]] loc0 : f32,
+ locations : FragLocations,
+ builtins : FragBuiltins) {
+ var col : f32 = ((builtins.coord.x * locations.loc1) + loc0);
+}
+)";
+
+ auto* expect = R"(
+struct FragBuiltins {
+ coord : vec4<f32>;
+};
+
+struct FragLocations {
+ loc1 : f32;
+ loc2 : vec4<u32>;
+};
+
+struct tint_symbol_2 {
+ [[location(0)]]
+ loc0 : f32;
+ [[location(1)]]
+ loc1 : f32;
+ [[location(2)]]
+ loc2 : vec4<u32>;
+};
+
+[[stage(fragment)]]
+fn frag_main([[builtin(position)]] tint_symbol_1 : vec4<f32>, tint_symbol : tint_symbol_2) {
+ let loc0 : f32 = tint_symbol.loc0;
+ let locations : FragLocations = FragLocations(tint_symbol.loc1, tint_symbol.loc2);
+ let builtins : FragBuiltins = FragBuiltins(tint_symbol_1);
+ var col : f32 = ((builtins.coord.x * locations.loc1) + loc0);
+}
+)";
+
+ DataMap data;
+ data.Add<CanonicalizeEntryPointIO::Config>(
+ CanonicalizeEntryPointIO::BuiltinStyle::kParameter);
+ auto got = Run<CanonicalizeEntryPointIO>(src, data);
+
+ EXPECT_EQ(expect, str(got));
+}
+
+TEST_F(CanonicalizeEntryPointIOTest, StructParameters_BuiltinsAsStructMembers) {
auto* src = R"(
struct FragBuiltins {
[[builtin(position)]] coord : vec4<f32>;
@@ -166,7 +307,10 @@
}
)";
- auto got = Run<CanonicalizeEntryPointIO>(src);
+ DataMap data;
+ data.Add<CanonicalizeEntryPointIO::Config>(
+ CanonicalizeEntryPointIO::BuiltinStyle::kStructMember);
+ auto got = Run<CanonicalizeEntryPointIO>(src, data);
EXPECT_EQ(expect, str(got));
}
@@ -191,7 +335,10 @@
}
)";
- auto got = Run<CanonicalizeEntryPointIO>(src);
+ DataMap data;
+ data.Add<CanonicalizeEntryPointIO::Config>(
+ CanonicalizeEntryPointIO::BuiltinStyle::kParameter);
+ auto got = Run<CanonicalizeEntryPointIO>(src, data);
EXPECT_EQ(expect, str(got));
}
@@ -240,7 +387,10 @@
}
)";
- auto got = Run<CanonicalizeEntryPointIO>(src);
+ DataMap data;
+ data.Add<CanonicalizeEntryPointIO::Config>(
+ CanonicalizeEntryPointIO::BuiltinStyle::kParameter);
+ auto got = Run<CanonicalizeEntryPointIO>(src, data);
EXPECT_EQ(expect, str(got));
}
@@ -304,7 +454,10 @@
}
)";
- auto got = Run<CanonicalizeEntryPointIO>(src);
+ DataMap data;
+ data.Add<CanonicalizeEntryPointIO::Config>(
+ CanonicalizeEntryPointIO::BuiltinStyle::kParameter);
+ auto got = Run<CanonicalizeEntryPointIO>(src, data);
EXPECT_EQ(expect, str(got));
}
@@ -366,7 +519,10 @@
}
)";
- auto got = Run<CanonicalizeEntryPointIO>(src);
+ DataMap data;
+ data.Add<CanonicalizeEntryPointIO::Config>(
+ CanonicalizeEntryPointIO::BuiltinStyle::kParameter);
+ auto got = Run<CanonicalizeEntryPointIO>(src, data);
EXPECT_EQ(expect, str(got));
}
@@ -444,7 +600,10 @@
}
)";
- auto got = Run<CanonicalizeEntryPointIO>(src);
+ DataMap data;
+ data.Add<CanonicalizeEntryPointIO::Config>(
+ CanonicalizeEntryPointIO::BuiltinStyle::kParameter);
+ auto got = Run<CanonicalizeEntryPointIO>(src, data);
EXPECT_EQ(expect, str(got));
}
@@ -501,7 +660,10 @@
}
)";
- auto got = Run<CanonicalizeEntryPointIO>(src);
+ DataMap data;
+ data.Add<CanonicalizeEntryPointIO::Config>(
+ CanonicalizeEntryPointIO::BuiltinStyle::kStructMember);
+ auto got = Run<CanonicalizeEntryPointIO>(src, data);
EXPECT_EQ(expect, str(got));
}
@@ -593,7 +755,10 @@
}
)";
- auto got = Run<CanonicalizeEntryPointIO>(src);
+ DataMap data;
+ data.Add<CanonicalizeEntryPointIO::Config>(
+ CanonicalizeEntryPointIO::BuiltinStyle::kStructMember);
+ auto got = Run<CanonicalizeEntryPointIO>(src, data);
EXPECT_EQ(expect, str(got));
}
@@ -617,7 +782,10 @@
}
)";
- auto got = Run<CanonicalizeEntryPointIO>(src);
+ DataMap data;
+ data.Add<CanonicalizeEntryPointIO::Config>(
+ CanonicalizeEntryPointIO::BuiltinStyle::kParameter);
+ auto got = Run<CanonicalizeEntryPointIO>(src, data);
EXPECT_EQ(expect, str(got));
}
diff --git a/src/transform/hlsl.cc b/src/transform/hlsl.cc
index 407e6b0..8579e49 100644
--- a/src/transform/hlsl.cc
+++ b/src/transform/hlsl.cc
@@ -32,8 +32,9 @@
Hlsl::Hlsl() = default;
Hlsl::~Hlsl() = default;
-Output Hlsl::Run(const Program* in, const DataMap& data) {
+Output Hlsl::Run(const Program* in, const DataMap&) {
Manager manager;
+ DataMap data;
manager.Add<CanonicalizeEntryPointIO>();
manager.Add<InlinePointerLets>();
// Simplify cleans up messy `*(&(expr))` expressions from InlinePointerLets.
@@ -46,6 +47,8 @@
manager.Add<CalculateArrayLength>();
manager.Add<ExternalTextureTransform>();
manager.Add<PromoteInitializersToConstVar>();
+ data.Add<CanonicalizeEntryPointIO::Config>(
+ CanonicalizeEntryPointIO::BuiltinStyle::kStructMember);
auto out = manager.Run(in, data);
if (!out.program.IsValid()) {
return out;
diff --git a/src/transform/msl.cc b/src/transform/msl.cc
index e8ae718..1d793b5 100644
--- a/src/transform/msl.cc
+++ b/src/transform/msl.cc
@@ -35,11 +35,14 @@
Msl::Msl() = default;
Msl::~Msl() = default;
-Output Msl::Run(const Program* in, const DataMap& data) {
+Output Msl::Run(const Program* in, const DataMap&) {
Manager manager;
+ DataMap data;
manager.Add<CanonicalizeEntryPointIO>();
manager.Add<ExternalTextureTransform>();
manager.Add<PromoteInitializersToConstVar>();
+ data.Add<CanonicalizeEntryPointIO::Config>(
+ CanonicalizeEntryPointIO::BuiltinStyle::kParameter);
auto out = manager.Run(in, data);
if (!out.program.IsValid()) {
return out;
diff --git a/src/writer/msl/generator_impl_function_test.cc b/src/writer/msl/generator_impl_function_test.cc
index 038caf6..d3641f1 100644
--- a/src/writer/msl/generator_impl_function_test.cc
+++ b/src/writer/msl/generator_impl_function_test.cc
@@ -140,16 +140,12 @@
using namespace metal;
struct tint_symbol_1 {
- float4 coord [[position]];
-};
-struct tint_symbol_2 {
float value [[depth(any)]];
};
-fragment tint_symbol_2 frag_main(tint_symbol_1 tint_symbol [[stage_in]]) {
- float4 const coord = tint_symbol.coord;
- tint_symbol_2 const tint_symbol_3 = {.value=coord.x};
- return tint_symbol_3;
+fragment tint_symbol_1 frag_main(float4 coord [[position]]) {
+ tint_symbol_1 const tint_symbol_2 = {.value=coord.x};
+ return tint_symbol_2;
}
)");
@@ -209,20 +205,19 @@
float col2 [[user(locn2)]];
float4 pos [[position]];
};
-struct tint_symbol_3 {
+struct tint_symbol_4 {
float col1 [[user(locn1)]];
float col2 [[user(locn2)]];
- float4 pos [[position]];
};
vertex tint_symbol vert_main() {
Interface const tint_symbol_1 = {.col1=0.5f, .col2=0.25f, .pos=float4()};
- tint_symbol const tint_symbol_4 = {.col1=tint_symbol_1.col1, .col2=tint_symbol_1.col2, .pos=tint_symbol_1.pos};
- return tint_symbol_4;
+ tint_symbol const tint_symbol_5 = {.col1=tint_symbol_1.col1, .col2=tint_symbol_1.col2, .pos=tint_symbol_1.pos};
+ return tint_symbol_5;
}
-fragment void frag_main(tint_symbol_3 tint_symbol_2 [[stage_in]]) {
- Interface const colors = {.col1=tint_symbol_2.col1, .col2=tint_symbol_2.col2, .pos=tint_symbol_2.pos};
+fragment void frag_main(float4 tint_symbol_3 [[position]], tint_symbol_4 tint_symbol_2 [[stage_in]]) {
+ Interface const colors = {.col1=tint_symbol_2.col1, .col2=tint_symbol_2.col2, .pos=tint_symbol_3};
float const r = colors.col1;
float const g = colors.col2;
return;
diff --git a/test/bug/tint/744.wgsl.expected.msl b/test/bug/tint/744.wgsl.expected.msl
index 6c0a282..66a945b 100644
--- a/test/bug/tint/744.wgsl.expected.msl
+++ b/test/bug/tint/744.wgsl.expected.msl
@@ -1,4 +1,4 @@
-SKIP: crbug.com/tint/817
+SKIP: crbug.com/tint/833 loop emission is broken
#include <metal_stdlib>
diff --git a/test/bug/tint/824.wgsl.expected.msl b/test/bug/tint/824.wgsl.expected.msl
index 44d2205..3489d64 100644
--- a/test/bug/tint/824.wgsl.expected.msl
+++ b/test/bug/tint/824.wgsl.expected.msl
@@ -1,5 +1,3 @@
-SKIP: crbug.com/tint/817
-
#include <metal_stdlib>
using namespace metal;
@@ -8,23 +6,24 @@
float4 color;
};
struct tint_symbol_2 {
- uint VertexIndex [[vertex_id]];
- uint InstanceIndex [[instance_id]];
-};
-struct tint_symbol_3 {
float4 color [[user(locn0)]];
float4 Position [[position]];
};
+struct tint_array_wrapper_0 {
+ float2 array[4];
+};
+struct tint_array_wrapper_1 {
+ float4 array[4];
+};
-vertex tint_symbol_3 tint_symbol(tint_symbol_2 tint_symbol_1 [[stage_in]]) {
- uint const VertexIndex = tint_symbol_1.VertexIndex;
- uint const InstanceIndex = tint_symbol_1.InstanceIndex;
- float2 const zv[4] = {float2(0.200000003f, 0.200000003f), float2(0.300000012f, 0.300000012f), float2(-0.100000001f, -0.100000001f), float2(1.100000024f, 1.100000024f)};
- float const z = zv[InstanceIndex].x;
+vertex tint_symbol_2 tint_symbol(uint VertexIndex [[vertex_id]], uint InstanceIndex [[instance_id]]) {
+ tint_array_wrapper_0 const zv = {float2(0.200000003f, 0.200000003f), float2(0.300000012f, 0.300000012f), float2(-0.100000001f, -0.100000001f), float2(1.100000024f, 1.100000024f)};
+ float const z = zv.array[InstanceIndex].x;
Output output = {};
output.Position = float4(0.5f, 0.5f, z, 1.0f);
- float4 const colors[4] = {float4(1.0f, 0.0f, 0.0f, 1.0f), float4(0.0f, 1.0f, 0.0f, 1.0f), float4(0.0f, 0.0f, 1.0f, 1.0f), float4(1.0f, 1.0f, 1.0f, 1.0f)};
- output.color = colors[InstanceIndex];
- return {output.color, output.Position};
+ tint_array_wrapper_1 const colors = {float4(1.0f, 0.0f, 0.0f, 1.0f), float4(0.0f, 1.0f, 0.0f, 1.0f), float4(0.0f, 0.0f, 1.0f, 1.0f), float4(1.0f, 1.0f, 1.0f, 1.0f)};
+ output.color = colors.array[InstanceIndex];
+ tint_symbol_2 const tint_symbol_3 = {.color=output.color, .Position=output.Position};
+ return tint_symbol_3;
}
diff --git a/test/samples/compute_boids.wgsl.expected.msl b/test/samples/compute_boids.wgsl.expected.msl
index 23298d4..22cc5da 100644
--- a/test/samples/compute_boids.wgsl.expected.msl
+++ b/test/samples/compute_boids.wgsl.expected.msl
@@ -1,5 +1,3 @@
-SKIP: crbug.com/tint/817
-
#include <metal_stdlib>
using namespace metal;
@@ -27,11 +25,11 @@
/* 0x0014 */ float rule2Scale;
/* 0x0018 */ float rule3Scale;
};
-struct Particles {
- /* 0x0000 */ Particle particles[5];
+struct tint_array_wrapper_0 {
+ Particle array[5];
};
-struct tint_symbol_5 {
- uint3 gl_GlobalInvocationID [[thread_position_in_grid]];
+struct Particles {
+ /* 0x0000 */ tint_array_wrapper_0 particles;
};
vertex tint_symbol_2 vert_main(tint_symbol_1 tint_symbol [[stage_in]]) {
@@ -40,21 +38,22 @@
float2 const a_pos = tint_symbol.a_pos;
float angle = -( atan2(a_particleVel.x, a_particleVel.y));
float2 pos = float2(((a_pos.x * cos(angle)) - (a_pos.y * sin(angle))), ((a_pos.x * sin(angle)) + (a_pos.y * cos(angle))));
- return {float4((pos + a_particlePos), 0.0f, 1.0f)};
+ tint_symbol_2 const tint_symbol_5 = {.value=float4((pos + a_particlePos), 0.0f, 1.0f)};
+ return tint_symbol_5;
}
fragment tint_symbol_3 frag_main() {
- return {float4(1.0f, 1.0f, 1.0f, 1.0f)};
+ tint_symbol_3 const tint_symbol_6 = {.value=float4(1.0f, 1.0f, 1.0f, 1.0f)};
+ return tint_symbol_6;
}
-kernel void comp_main(tint_symbol_5 tint_symbol_4 [[stage_in]], constant SimParams& params [[buffer(0)]], device Particles& particlesA [[buffer(1)]], device Particles& particlesB [[buffer(2)]]) {
- uint3 const gl_GlobalInvocationID = tint_symbol_4.gl_GlobalInvocationID;
+kernel void comp_main(uint3 gl_GlobalInvocationID [[thread_position_in_grid]], constant SimParams& params [[buffer(0)]], device Particles& particlesA [[buffer(1)]], device Particles& particlesB [[buffer(2)]]) {
uint index = gl_GlobalInvocationID.x;
if ((index >= 5u)) {
return;
}
- float2 vPos = particlesA.particles[index].pos;
- float2 vVel = particlesA.particles[index].vel;
+ float2 vPos = particlesA.particles.array[index].pos;
+ float2 vVel = particlesA.particles.array[index].vel;
float2 cMass = float2(0.0f, 0.0f);
float2 cVel = float2(0.0f, 0.0f);
float2 colVel = float2(0.0f, 0.0f);
@@ -78,8 +77,8 @@
if ((i == index)) {
continue;
}
- pos = particlesA.particles[i].pos.xy;
- vel = particlesA.particles[i].vel.xy;
+ pos = particlesA.particles.array[i].pos.xy;
+ vel = particlesA.particles.array[i].vel.xy;
if (( distance(pos, vPos) < params.rule1Distance)) {
cMass = (cMass + pos);
cMassCount = (cMassCount + 1);
@@ -115,8 +114,8 @@
if ((vPos.y > 1.0f)) {
vPos.y = -1.0f;
}
- particlesB.particles[index].pos = vPos;
- particlesB.particles[index].vel = vVel;
+ particlesB.particles.array[index].pos = vPos;
+ particlesB.particles.array[index].vel = vVel;
return;
}
diff --git a/test/samples/triangle.wgsl.expected.msl b/test/samples/triangle.wgsl.expected.msl
index 3d108db..7f4a0ad 100644
--- a/test/samples/triangle.wgsl.expected.msl
+++ b/test/samples/triangle.wgsl.expected.msl
@@ -1,25 +1,24 @@
-SKIP: crbug.com/tint/817
-
#include <metal_stdlib>
using namespace metal;
struct tint_symbol_1 {
- int VertexIndex [[vertex_id]];
-};
-struct tint_symbol_2 {
float4 value [[position]];
};
-struct tint_symbol_3 {
+struct tint_symbol_2 {
float4 value [[color(0)]];
};
+struct tint_array_wrapper_0 {
+ float2 array[3];
+};
-constant float2 pos[3] = {float2(0.0f, 0.5f), float2(-0.5f, -0.5f), float2(0.5f, -0.5f)};
-vertex tint_symbol_2 vtx_main(tint_symbol_1 tint_symbol [[stage_in]]) {
- int const VertexIndex = tint_symbol.VertexIndex;
- return {float4(pos[VertexIndex], 0.0f, 1.0f)};
+constant tint_array_wrapper_0 pos = {float2(0.0f, 0.5f), float2(-0.5f, -0.5f), float2(0.5f, -0.5f)};
+vertex tint_symbol_1 vtx_main(uint VertexIndex [[vertex_id]]) {
+ tint_symbol_1 const tint_symbol_3 = {.value=float4(pos.array[VertexIndex], 0.0f, 1.0f)};
+ return tint_symbol_3;
}
-fragment tint_symbol_3 frag_main() {
- return {float4(1.0f, 0.0f, 0.0f, 1.0f)};
+fragment tint_symbol_2 frag_main() {
+ tint_symbol_2 const tint_symbol_4 = {.value=float4(1.0f, 0.0f, 0.0f, 1.0f)};
+ return tint_symbol_4;
}
diff --git a/test/shader_io/compute_input_builtins.wgsl.expected.msl b/test/shader_io/compute_input_builtins.wgsl.expected.msl
index 71d62fb..2774a3d 100644
--- a/test/shader_io/compute_input_builtins.wgsl.expected.msl
+++ b/test/shader_io/compute_input_builtins.wgsl.expected.msl
@@ -1 +1,8 @@
-SKIP: crbug.com/tint/817 attribute only applies to parameters
+#include <metal_stdlib>
+
+using namespace metal;
+kernel void tint_symbol(uint3 local_invocation_id [[thread_position_in_threadgroup]], uint local_invocation_index [[thread_index_in_threadgroup]], uint3 global_invocation_id [[thread_position_in_grid]], uint3 workgroup_id [[threadgroup_position_in_grid]]) {
+ uint const foo = (((local_invocation_id.x + local_invocation_index) + global_invocation_id.x) + workgroup_id.x);
+ return;
+}
+
diff --git a/test/shader_io/compute_input_builtins_struct.wgsl.expected.msl b/test/shader_io/compute_input_builtins_struct.wgsl.expected.msl
index 71d62fb..9a16f1b 100644
--- a/test/shader_io/compute_input_builtins_struct.wgsl.expected.msl
+++ b/test/shader_io/compute_input_builtins_struct.wgsl.expected.msl
@@ -1 +1,16 @@
-SKIP: crbug.com/tint/817 attribute only applies to parameters
+#include <metal_stdlib>
+
+using namespace metal;
+struct ComputeInputs {
+ uint3 local_invocation_id;
+ uint local_invocation_index;
+ uint3 global_invocation_id;
+ uint3 workgroup_id;
+};
+
+kernel void tint_symbol(uint3 tint_symbol_2 [[thread_position_in_threadgroup]], uint tint_symbol_3 [[thread_index_in_threadgroup]], uint3 tint_symbol_4 [[thread_position_in_grid]], uint3 tint_symbol_5 [[threadgroup_position_in_grid]]) {
+ ComputeInputs const inputs = {.local_invocation_id=tint_symbol_2, .local_invocation_index=tint_symbol_3, .global_invocation_id=tint_symbol_4, .workgroup_id=tint_symbol_5};
+ uint const foo = (((inputs.local_invocation_id.x + inputs.local_invocation_index) + inputs.global_invocation_id.x) + inputs.workgroup_id.x);
+ return;
+}
+
diff --git a/test/shader_io/compute_input_mixed.wgsl.expected.msl b/test/shader_io/compute_input_mixed.wgsl.expected.msl
index 71d62fb..fe6293b 100644
--- a/test/shader_io/compute_input_mixed.wgsl.expected.msl
+++ b/test/shader_io/compute_input_mixed.wgsl.expected.msl
@@ -1 +1,17 @@
-SKIP: crbug.com/tint/817 attribute only applies to parameters
+#include <metal_stdlib>
+
+using namespace metal;
+struct ComputeInputs0 {
+ uint3 local_invocation_id;
+};
+struct ComputeInputs1 {
+ uint3 workgroup_id;
+};
+
+kernel void tint_symbol(uint3 tint_symbol_2 [[thread_position_in_threadgroup]], uint local_invocation_index [[thread_index_in_threadgroup]], uint3 global_invocation_id [[thread_position_in_grid]], uint3 tint_symbol_3 [[threadgroup_position_in_grid]]) {
+ ComputeInputs0 const inputs0 = {.local_invocation_id=tint_symbol_2};
+ ComputeInputs1 const inputs1 = {.workgroup_id=tint_symbol_3};
+ uint const foo = (((inputs0.local_invocation_id.x + local_invocation_index) + global_invocation_id.x) + inputs1.workgroup_id.x);
+ return;
+}
+
diff --git a/test/shader_io/fragment_input_builtins.wgsl.expected.msl b/test/shader_io/fragment_input_builtins.wgsl.expected.msl
index 3e46c6f..4204e5e 100644
--- a/test/shader_io/fragment_input_builtins.wgsl.expected.msl
+++ b/test/shader_io/fragment_input_builtins.wgsl.expected.msl
@@ -1,18 +1,7 @@
#include <metal_stdlib>
using namespace metal;
-struct tint_symbol_2 {
- float4 position [[position]];
- bool front_facing [[front_facing]];
- uint sample_index [[sample_id]];
- uint sample_mask [[sample_mask]];
-};
-
-fragment void tint_symbol(tint_symbol_2 tint_symbol_1 [[stage_in]]) {
- float4 const position = tint_symbol_1.position;
- bool const front_facing = tint_symbol_1.front_facing;
- uint const sample_index = tint_symbol_1.sample_index;
- uint const sample_mask = tint_symbol_1.sample_mask;
+fragment void tint_symbol(float4 position [[position]], bool front_facing [[front_facing]], uint sample_index [[sample_id]], uint sample_mask [[sample_mask]]) {
if (front_facing) {
float4 const foo = position;
uint const bar = (sample_index + sample_mask);
diff --git a/test/shader_io/fragment_input_builtins_struct.wgsl.expected.msl b/test/shader_io/fragment_input_builtins_struct.wgsl.expected.msl
index 7f05ccf..284dbe5 100644
--- a/test/shader_io/fragment_input_builtins_struct.wgsl.expected.msl
+++ b/test/shader_io/fragment_input_builtins_struct.wgsl.expected.msl
@@ -7,15 +7,9 @@
uint sample_index;
uint sample_mask;
};
-struct tint_symbol_2 {
- float4 position [[position]];
- bool front_facing [[front_facing]];
- uint sample_index [[sample_id]];
- uint sample_mask [[sample_mask]];
-};
-fragment void tint_symbol(tint_symbol_2 tint_symbol_1 [[stage_in]]) {
- FragmentInputs const inputs = {.position=tint_symbol_1.position, .front_facing=tint_symbol_1.front_facing, .sample_index=tint_symbol_1.sample_index, .sample_mask=tint_symbol_1.sample_mask};
+fragment void tint_symbol(float4 tint_symbol_2 [[position]], bool tint_symbol_3 [[front_facing]], uint tint_symbol_4 [[sample_id]], uint tint_symbol_5 [[sample_mask]]) {
+ FragmentInputs const inputs = {.position=tint_symbol_2, .front_facing=tint_symbol_3, .sample_index=tint_symbol_4, .sample_mask=tint_symbol_5};
if (inputs.front_facing) {
float4 const foo = inputs.position;
uint const bar = (inputs.sample_index + inputs.sample_mask);
diff --git a/test/shader_io/fragment_input_mixed.wgsl.expected.msl b/test/shader_io/fragment_input_mixed.wgsl.expected.msl
index 4e7aa7f..9b3a13a 100644
--- a/test/shader_io/fragment_input_mixed.wgsl.expected.msl
+++ b/test/shader_io/fragment_input_mixed.wgsl.expected.msl
@@ -9,23 +9,17 @@
float4 loc3;
uint sample_mask;
};
-struct tint_symbol_2 {
+struct tint_symbol_4 {
int loc0 [[user(locn0)]];
uint loc1 [[user(locn1)]];
float loc2 [[user(locn2)]];
float4 loc3 [[user(locn3)]];
- float4 position [[position]];
- bool front_facing [[front_facing]];
- uint sample_index [[sample_id]];
- uint sample_mask [[sample_mask]];
};
-fragment void tint_symbol(tint_symbol_2 tint_symbol_1 [[stage_in]]) {
- FragmentInputs0 const inputs0 = {.position=tint_symbol_1.position, .loc0=tint_symbol_1.loc0};
- bool const front_facing = tint_symbol_1.front_facing;
+fragment void tint_symbol(float4 tint_symbol_2 [[position]], bool front_facing [[front_facing]], uint sample_index [[sample_id]], uint tint_symbol_3 [[sample_mask]], tint_symbol_4 tint_symbol_1 [[stage_in]]) {
+ FragmentInputs0 const inputs0 = {.position=tint_symbol_2, .loc0=tint_symbol_1.loc0};
uint const loc1 = tint_symbol_1.loc1;
- uint const sample_index = tint_symbol_1.sample_index;
- FragmentInputs1 const inputs1 = {.loc3=tint_symbol_1.loc3, .sample_mask=tint_symbol_1.sample_mask};
+ FragmentInputs1 const inputs1 = {.loc3=tint_symbol_1.loc3, .sample_mask=tint_symbol_3};
float const loc2 = tint_symbol_1.loc2;
if (front_facing) {
float4 const foo = inputs0.position;
diff --git a/test/shader_io/shared_struct_different_stages.wgsl.expected.msl b/test/shader_io/shared_struct_different_stages.wgsl.expected.msl
index 7183f0a..f51829c 100644
--- a/test/shader_io/shared_struct_different_stages.wgsl.expected.msl
+++ b/test/shader_io/shared_struct_different_stages.wgsl.expected.msl
@@ -11,20 +11,19 @@
float col2 [[user(locn2)]];
float4 pos [[position]];
};
-struct tint_symbol_3 {
+struct tint_symbol_4 {
float col1 [[user(locn1)]];
float col2 [[user(locn2)]];
- float4 pos [[position]];
};
vertex tint_symbol vert_main() {
Interface const tint_symbol_1 = {.col1=0.400000006f, .col2=0.600000024f, .pos=float4()};
- tint_symbol const tint_symbol_4 = {.col1=tint_symbol_1.col1, .col2=tint_symbol_1.col2, .pos=tint_symbol_1.pos};
- return tint_symbol_4;
+ tint_symbol const tint_symbol_5 = {.col1=tint_symbol_1.col1, .col2=tint_symbol_1.col2, .pos=tint_symbol_1.pos};
+ return tint_symbol_5;
}
-fragment void frag_main(tint_symbol_3 tint_symbol_2 [[stage_in]]) {
- Interface const colors = {.col1=tint_symbol_2.col1, .col2=tint_symbol_2.col2, .pos=tint_symbol_2.pos};
+fragment void frag_main(float4 tint_symbol_3 [[position]], tint_symbol_4 tint_symbol_2 [[stage_in]]) {
+ Interface const colors = {.col1=tint_symbol_2.col1, .col2=tint_symbol_2.col2, .pos=tint_symbol_3};
float const r = colors.col1;
float const g = colors.col2;
return;
diff --git a/test/shader_io/shared_struct_storage_buffer.wgsl.expected.msl b/test/shader_io/shared_struct_storage_buffer.wgsl.expected.msl
index 63eeabf..63fe338 100644
--- a/test/shader_io/shared_struct_storage_buffer.wgsl.expected.msl
+++ b/test/shader_io/shared_struct_storage_buffer.wgsl.expected.msl
@@ -8,14 +8,13 @@
/* 0x0080 */ packed_float4 v;
/* 0x0090 */ int8_t tint_pad_1[112];
};
-struct tint_symbol_1 {
+struct tint_symbol_2 {
float f [[user(locn0)]];
uint u [[user(locn1)]];
- float4 v [[position]];
};
-fragment void frag_main(tint_symbol_1 tint_symbol [[stage_in]], device S& output [[buffer(0)]]) {
- S const input = {.f=tint_symbol.f, .u=tint_symbol.u, .v=tint_symbol.v};
+fragment void frag_main(float4 tint_symbol_1 [[position]], tint_symbol_2 tint_symbol [[stage_in]], device S& output [[buffer(0)]]) {
+ S const input = {.f=tint_symbol.f, .u=tint_symbol.u, .v=tint_symbol_1};
float const f = input.f;
uint const u = input.u;
float4 const v = input.v;
diff --git a/test/shader_io/vertex_input_builtins.wgsl.expected.msl b/test/shader_io/vertex_input_builtins.wgsl.expected.msl
index 71d62fb..f3cff19 100644
--- a/test/shader_io/vertex_input_builtins.wgsl.expected.msl
+++ b/test/shader_io/vertex_input_builtins.wgsl.expected.msl
@@ -1 +1,13 @@
-SKIP: crbug.com/tint/817 attribute only applies to parameters
+#include <metal_stdlib>
+
+using namespace metal;
+struct tint_symbol_2 {
+ float4 value [[position]];
+};
+
+vertex tint_symbol_2 tint_symbol(uint vertex_index [[vertex_id]], uint instance_index [[instance_id]]) {
+ uint const foo = (vertex_index + instance_index);
+ tint_symbol_2 const tint_symbol_3 = {.value=float4()};
+ return tint_symbol_3;
+}
+
diff --git a/test/shader_io/vertex_input_builtins_struct.wgsl.expected.msl b/test/shader_io/vertex_input_builtins_struct.wgsl.expected.msl
index 71d62fb..c973c83 100644
--- a/test/shader_io/vertex_input_builtins_struct.wgsl.expected.msl
+++ b/test/shader_io/vertex_input_builtins_struct.wgsl.expected.msl
@@ -1 +1,18 @@
-SKIP: crbug.com/tint/817 attribute only applies to parameters
+#include <metal_stdlib>
+
+using namespace metal;
+struct VertexInputs {
+ uint vertex_index;
+ uint instance_index;
+};
+struct tint_symbol_4 {
+ float4 value [[position]];
+};
+
+vertex tint_symbol_4 tint_symbol(uint tint_symbol_2 [[vertex_id]], uint tint_symbol_3 [[instance_id]]) {
+ VertexInputs const inputs = {.vertex_index=tint_symbol_2, .instance_index=tint_symbol_3};
+ uint const foo = (inputs.vertex_index + inputs.instance_index);
+ tint_symbol_4 const tint_symbol_5 = {.value=float4()};
+ return tint_symbol_5;
+}
+
diff --git a/test/shader_io/vertex_input_mixed.wgsl.expected.msl b/test/shader_io/vertex_input_mixed.wgsl.expected.msl
index 71d62fb..9c1e813 100644
--- a/test/shader_io/vertex_input_mixed.wgsl.expected.msl
+++ b/test/shader_io/vertex_input_mixed.wgsl.expected.msl
@@ -1 +1,34 @@
-SKIP: crbug.com/tint/817 attribute only applies to parameters
+#include <metal_stdlib>
+
+using namespace metal;
+struct VertexInputs0 {
+ uint vertex_index;
+ int loc0;
+};
+struct VertexInputs1 {
+ float loc2;
+ float4 loc3;
+};
+struct tint_symbol_3 {
+ int loc0 [[attribute(0)]];
+ uint loc1 [[attribute(1)]];
+ float loc2 [[attribute(2)]];
+ float4 loc3 [[attribute(3)]];
+};
+struct tint_symbol_4 {
+ float4 value [[position]];
+};
+
+vertex tint_symbol_4 tint_symbol(uint tint_symbol_2 [[vertex_id]], uint instance_index [[instance_id]], tint_symbol_3 tint_symbol_1 [[stage_in]]) {
+ VertexInputs0 const inputs0 = {.vertex_index=tint_symbol_2, .loc0=tint_symbol_1.loc0};
+ uint const loc1 = tint_symbol_1.loc1;
+ VertexInputs1 const inputs1 = {.loc2=tint_symbol_1.loc2, .loc3=tint_symbol_1.loc3};
+ uint const foo = (inputs0.vertex_index + instance_index);
+ int const i = inputs0.loc0;
+ uint const u = loc1;
+ float const f = inputs1.loc2;
+ float4 const v = inputs1.loc3;
+ tint_symbol_4 const tint_symbol_5 = {.value=float4()};
+ return tint_symbol_5;
+}
+
