[msl] Bump default version to 2.2
Our minimum supported macOS version is 10.15, which corresponds to
Metal 2.2.
This fixes a few E2E tests that used nullptr for unused module-scope
variables.
Change-Id: I112c88fc9139ef9806a9fb52501b3c6c19754f19
Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/214234
Reviewed-by: dan sinclair <dsinclair@chromium.org>
Auto-Submit: James Price <jrprice@google.com>
Commit-Queue: dan sinclair <dsinclair@chromium.org>
diff --git a/src/tint/cmd/remote_compile/main.cc b/src/tint/cmd/remote_compile/main.cc
index 277450e..7ac73ae 100644
--- a/src/tint/cmd/remote_compile/main.cc
+++ b/src/tint/cmd/remote_compile/main.cc
@@ -450,10 +450,7 @@
}
#if TINT_BUILD_MSL_WRITER && TINT_BUILD_IS_MAC
if (req.language == SourceLanguage::MSL) {
- auto version = tint::msl::validate::MslVersion::kMsl_1_2;
- if (req.version_major == 2 && req.version_minor == 1) {
- version = tint::msl::validate::MslVersion::kMsl_2_1;
- }
+ auto version = tint::msl::validate::MslVersion::kMsl_2_2;
if (req.version_major == 2 && req.version_minor == 3) {
version = tint::msl::validate::MslVersion::kMsl_2_3;
}
diff --git a/src/tint/cmd/tint/main.cc b/src/tint/cmd/tint/main.cc
index f9403bd..7925c02 100644
--- a/src/tint/cmd/tint/main.cc
+++ b/src/tint/cmd/tint/main.cc
@@ -866,14 +866,10 @@
PrintHash(hash);
}
- // Default to validating against MSL 1.2.
- // If subgroups are used, bump the version to 2.2.
- auto msl_version = tint::msl::validate::MslVersion::kMsl_1_2;
+ // Default to validating against MSL 2.2, which corresponds to macOS 10.15.
+ // Check for extensions that bump the requirements.
+ auto msl_version = tint::msl::validate::MslVersion::kMsl_2_2;
for (auto* enable : program.AST().Enables()) {
- if (enable->HasExtension(tint::wgsl::Extension::kChromiumExperimentalSubgroups) ||
- enable->HasExtension(tint::wgsl::Extension::kSubgroups)) {
- msl_version = std::max(msl_version, tint::msl::validate::MslVersion::kMsl_2_2);
- }
if (enable->HasExtension(tint::wgsl::Extension::kChromiumExperimentalPixelLocal) ||
enable->HasExtension(tint::wgsl::Extension::kChromiumExperimentalFramebufferFetch)) {
msl_version = std::max(msl_version, tint::msl::validate::MslVersion::kMsl_2_3);
diff --git a/src/tint/lang/msl/validate/validate.cc b/src/tint/lang/msl/validate/validate.cc
index 3b75532..05dee69 100644
--- a/src/tint/lang/msl/validate/validate.cc
+++ b/src/tint/lang/msl/validate/validate.cc
@@ -47,12 +47,6 @@
const char* version_str = nullptr;
switch (version) {
- case MslVersion::kMsl_1_2:
- version_str = "-std=macos-metal1.2";
- break;
- case MslVersion::kMsl_2_1:
- version_str = "-std=macos-metal2.1";
- break;
case MslVersion::kMsl_2_2:
version_str = "-std=macos-metal2.2";
break;
diff --git a/src/tint/lang/msl/validate/validate.h b/src/tint/lang/msl/validate/validate.h
index f1dfb92..167aa52 100644
--- a/src/tint/lang/msl/validate/validate.h
+++ b/src/tint/lang/msl/validate/validate.h
@@ -36,8 +36,6 @@
/// The version of MSL to validate against.
/// Note: these must kept be in ascending order
enum class MslVersion {
- kMsl_1_2,
- kMsl_2_1,
kMsl_2_2,
kMsl_2_3,
};
diff --git a/src/tint/lang/msl/validate/validate_metal.mm b/src/tint/lang/msl/validate/validate_metal.mm
index 0d5d507..318d0f4 100644
--- a/src/tint/lang/msl/validate/validate_metal.mm
+++ b/src/tint/lang/msl/validate/validate_metal.mm
@@ -50,12 +50,6 @@
MTLCompileOptions* compileOptions = [MTLCompileOptions new];
compileOptions.fastMathEnabled = true;
switch (version) {
- case MslVersion::kMsl_1_2:
- compileOptions.languageVersion = MTLLanguageVersion1_2;
- break;
- case MslVersion::kMsl_2_1:
- compileOptions.languageVersion = MTLLanguageVersion2_1;
- break;
case MslVersion::kMsl_2_2:
compileOptions.languageVersion = MTLLanguageVersion2_2;
break;
diff --git a/test/tint/benchmark/particles.wgsl.expected.ir.msl b/test/tint/benchmark/particles.wgsl.expected.ir.msl
deleted file mode 100644
index 5237630..0000000
--- a/test/tint/benchmark/particles.wgsl.expected.ir.msl
+++ /dev/null
@@ -1,254 +0,0 @@
-SKIP: FAILED
-
-#include <metal_stdlib>
-using namespace metal;
-
-struct RenderParams {
- float4x4 modelViewProjectionMatrix;
- float3 right;
- float3 up;
-};
-
-struct SimulationParams {
- float deltaTime;
- float4 seed;
-};
-
-template<typename T, size_t N>
-struct tint_array {
- const constant T& operator[](size_t i) const constant { return elements[i]; }
- device T& operator[](size_t i) device { return elements[i]; }
- const device T& operator[](size_t i) const device { return elements[i]; }
- thread T& operator[](size_t i) thread { return elements[i]; }
- const thread T& operator[](size_t i) const thread { return elements[i]; }
- threadgroup T& operator[](size_t i) threadgroup { return elements[i]; }
- const threadgroup T& operator[](size_t i) const threadgroup { return elements[i]; }
- T elements[N];
-};
-
-struct Particle {
- float3 position;
- float lifetime;
- float4 color;
- float3 velocity;
-};
-
-struct Particles {
- tint_array<Particle, 1> particles;
-};
-
-struct UBO {
- uint width;
-};
-
-struct Buffer {
- tint_array<float, 1> weights;
-};
-
-struct tint_module_vars_struct {
- thread float2* rand_seed;
- const constant RenderParams* render_params;
- const constant SimulationParams* sim_params;
- device Particles* data;
- texture2d<float, access::sample> tint_symbol;
- const constant UBO* ubo;
- const device Buffer* buf_in;
- device Buffer* buf_out;
- texture2d<float, access::sample> tex_in;
- texture2d<float, access::write> tex_out;
-};
-
-struct VertexOutput {
- float4 position;
- float4 color;
- float2 quad_pos;
-};
-
-struct VertexInput {
- float3 position;
- float4 color;
- float2 quad_pos;
-};
-
-struct vs_main_outputs {
- float4 VertexOutput_position [[position]];
- float4 VertexOutput_color [[user(locn0)]];
- float2 VertexOutput_quad_pos [[user(locn1)]];
-};
-
-struct vs_main_inputs {
- float3 VertexInput_position [[attribute(0)]];
- float4 VertexInput_color [[attribute(1)]];
- float2 VertexInput_quad_pos [[attribute(2)]];
-};
-
-struct fs_main_outputs {
- float4 tint_symbol_1 [[color(0)]];
-};
-
-struct fs_main_inputs {
- float4 VertexOutput_color [[user(locn0)]];
- float2 VertexOutput_quad_pos [[user(locn1)]];
-};
-
-float rand(tint_module_vars_struct tint_module_vars) {
- (*tint_module_vars.rand_seed)[0u] = fract((cos(dot((*tint_module_vars.rand_seed), float2(23.1407794952392578125f, 232.6168975830078125f))) * 136.816802978515625f));
- (*tint_module_vars.rand_seed)[1u] = fract((cos(dot((*tint_module_vars.rand_seed), float2(54.478565216064453125f, 345.841522216796875f))) * 534.7645263671875f));
- return (*tint_module_vars.rand_seed)[1u];
-}
-
-VertexOutput vs_main_inner(VertexInput in, tint_module_vars_struct tint_module_vars) {
- float3 quad_pos = (float2x3((*tint_module_vars.render_params).right, (*tint_module_vars.render_params).up) * in.quad_pos);
- float3 position = (in.position + (quad_pos * 0.00999999977648258209f));
- VertexOutput out = {};
- float4x4 const v = (*tint_module_vars.render_params).modelViewProjectionMatrix;
- out.position = (v * float4(position, 1.0f));
- out.color = in.color;
- out.quad_pos = in.quad_pos;
- return out;
-}
-
-float4 fs_main_inner(VertexOutput in) {
- float4 color = in.color;
- float const v_1 = color[3u];
- color[3u] = (v_1 * max((1.0f - length(in.quad_pos)), 0.0f));
- return color;
-}
-
-void tint_store_and_preserve_padding(device Particle* const target, Particle value_param) {
- (*target).position = value_param.position;
- (*target).lifetime = value_param.lifetime;
- (*target).color = value_param.color;
- (*target).velocity = value_param.velocity;
-}
-
-void simulate_inner(uint3 GlobalInvocationID, tint_module_vars_struct tint_module_vars) {
- float2 const v_2 = (*tint_module_vars.sim_params).seed.xy;
- float2 const v_3 = (v_2 + float2(GlobalInvocationID.xy));
- (*tint_module_vars.rand_seed) = (v_3 * (*tint_module_vars.sim_params).seed.zw);
- uint const idx = GlobalInvocationID[0u];
- Particle particle = (*tint_module_vars.data).particles[idx];
- particle.velocity[2u] = (particle.velocity[2u] - ((*tint_module_vars.sim_params).deltaTime * 0.5f));
- particle.position = (particle.position + ((*tint_module_vars.sim_params).deltaTime * particle.velocity));
- particle.lifetime = (particle.lifetime - (*tint_module_vars.sim_params).deltaTime);
- particle.color[3u] = smoothstep(0.0f, 0.5f, particle.lifetime);
- if ((particle.lifetime < 0.0f)) {
- uint2 coord = uint2(0u);
- {
- uint level = (tint_module_vars.tint_symbol.get_num_mip_levels() - 1u);
- while(true) {
- if ((level > 0u)) {
- } else {
- break;
- }
- float4 const probabilites = tint_module_vars.tint_symbol.read(coord, level);
- float4 const value = float4(rand(tint_module_vars));
- bool4 const mask = ((value >= float4(0.0f, probabilites.xyz)) & (value < probabilites));
- coord = (coord * 2u);
- uint const v_4 = select(0u, 1u, any(mask.yw));
- coord[0u] = (coord[0u] + v_4);
- uint const v_5 = select(0u, 1u, any(mask.zw));
- coord[1u] = (coord[1u] + v_5);
- {
- level = (level - 1u);
- }
- continue;
- }
- }
- float2 const v_6 = float2(coord);
- uint const v_7 = tint_module_vars.tint_symbol.get_width(0u);
- float2 const uv = (v_6 / float2(uint2(v_7, tint_module_vars.tint_symbol.get_height(0u))));
- particle.position = float3((((uv - 0.5f) * 3.0f) * float2(1.0f, -1.0f)), 0.0f);
- particle.color = tint_module_vars.tint_symbol.read(coord, 0);
- particle.velocity[0u] = ((rand(tint_module_vars) - 0.5f) * 0.10000000149011611938f);
- particle.velocity[1u] = ((rand(tint_module_vars) - 0.5f) * 0.10000000149011611938f);
- particle.velocity[2u] = (rand(tint_module_vars) * 0.30000001192092895508f);
- particle.lifetime = (0.5f + (rand(tint_module_vars) * 2.0f));
- }
- tint_store_and_preserve_padding((&(*tint_module_vars.data).particles[idx]), particle);
-}
-
-void import_level_inner(uint3 coord, tint_module_vars_struct tint_module_vars) {
- uint const offset = (coord[0u] + (coord[1u] * (*tint_module_vars.ubo).width));
- (*tint_module_vars.buf_out).weights[offset] = tint_module_vars.tex_in.read(uint2(int2(coord.xy)), 0)[3u];
-}
-
-void export_level_inner(uint3 coord, tint_module_vars_struct tint_module_vars) {
- uint const v_8 = tint_module_vars.tex_out.get_width(0u);
- if (all((coord.xy < uint2(uint2(v_8, tint_module_vars.tex_out.get_height(0u)))))) {
- uint const dst_offset = (coord[0u] + (coord[1u] * (*tint_module_vars.ubo).width));
- uint const src_offset = ((coord[0u] * 2u) + ((coord[1u] * 2u) * (*tint_module_vars.ubo).width));
- float const a = (*tint_module_vars.buf_in).weights[(src_offset + 0u)];
- float const b = (*tint_module_vars.buf_in).weights[(src_offset + 1u)];
- float const c = (*tint_module_vars.buf_in).weights[((src_offset + 0u) + (*tint_module_vars.ubo).width)];
- float const d = (*tint_module_vars.buf_in).weights[((src_offset + 1u) + (*tint_module_vars.ubo).width)];
- float const sum = dot(float4(a, b, c, d), float4(1.0f));
- (*tint_module_vars.buf_out).weights[dst_offset] = (sum / 4.0f);
- float4 const v_9 = float4(a, (a + b), ((a + b) + c), sum);
- float4 const probabilities = (v_9 / max(sum, 0.00009999999747378752f));
- tint_module_vars.tex_out.write(probabilities, uint2(int2(coord.xy)));
- }
-}
-
-vertex vs_main_outputs vs_main(vs_main_inputs inputs [[stage_in]], const constant RenderParams* render_params [[buffer(0)]]) {
- tint_module_vars_struct const tint_module_vars = tint_module_vars_struct{.render_params=render_params};
- VertexOutput const v_10 = vs_main_inner(VertexInput{.position=inputs.VertexInput_position, .color=inputs.VertexInput_color, .quad_pos=inputs.VertexInput_quad_pos}, tint_module_vars);
- return vs_main_outputs{.VertexOutput_position=v_10.position, .VertexOutput_color=v_10.color, .VertexOutput_quad_pos=v_10.quad_pos};
-}
-
-fragment fs_main_outputs fs_main(float4 VertexOutput_position [[position]], fs_main_inputs inputs [[stage_in]]) {
- return fs_main_outputs{.tint_symbol_1=fs_main_inner(VertexOutput{.position=VertexOutput_position, .color=inputs.VertexOutput_color, .quad_pos=inputs.VertexOutput_quad_pos})};
-}
-
-kernel void simulate(uint3 GlobalInvocationID [[thread_position_in_grid]], const constant SimulationParams* sim_params [[buffer(0)]], device Particles* data [[buffer(1)]], texture2d<float, access::sample> tint_symbol [[texture(0)]]) {
- thread float2 rand_seed = 0.0f;
- tint_module_vars_struct const tint_module_vars = tint_module_vars_struct{.rand_seed=(&rand_seed), .sim_params=sim_params, .data=data, .tint_symbol=tint_symbol};
- simulate_inner(GlobalInvocationID, tint_module_vars);
-}
-
-kernel void import_level(uint3 coord [[thread_position_in_grid]], const constant UBO* ubo [[buffer(2)]], device Buffer* buf_out [[buffer(3)]], texture2d<float, access::sample> tex_in [[texture(1)]]) {
- tint_module_vars_struct const tint_module_vars = tint_module_vars_struct{.ubo=ubo, .buf_out=buf_out, .tex_in=tex_in};
- import_level_inner(coord, tint_module_vars);
-}
-
-kernel void export_level(uint3 coord [[thread_position_in_grid]], const constant UBO* ubo [[buffer(2)]], const device Buffer* buf_in [[buffer(4)]], device Buffer* buf_out [[buffer(3)]], texture2d<float, access::write> tex_out [[texture(2)]]) {
- tint_module_vars_struct const tint_module_vars = tint_module_vars_struct{.ubo=ubo, .buf_in=buf_in, .buf_out=buf_out, .tex_out=tex_out};
- export_level_inner(coord, tint_module_vars);
-}
-program_source:192:104: error: call to deleted constructor of 'texture2d<float, access::sample>'
- tint_module_vars_struct const tint_module_vars = tint_module_vars_struct{.render_params=render_params};
- ^
-/System/Library/PrivateFrameworks/GPUCompiler.framework/Versions/32023/Libraries/lib/clang/32023.194/include/metal/metal_texture:6083:3: note: 'texture2d' has been explicitly marked deleted here
- texture2d() thread = delete;
- ^
-program_source:51:36: note: in implicit initialization of field 'tint_symbol' with omitted initializer
- texture2d<float, access::sample> tint_symbol;
- ^
-program_source:203:161: error: call to deleted constructor of 'texture2d<float, access::sample>'
- tint_module_vars_struct const tint_module_vars = tint_module_vars_struct{.rand_seed=(&rand_seed), .sim_params=sim_params, .data=data, .tint_symbol=tint_symbol};
- ^
-/System/Library/PrivateFrameworks/GPUCompiler.framework/Versions/32023/Libraries/lib/clang/32023.194/include/metal/metal_texture:6083:3: note: 'texture2d' has been explicitly marked deleted here
- texture2d() thread = delete;
- ^
-program_source:55:36: note: in implicit initialization of field 'tex_in' with omitted initializer
- texture2d<float, access::sample> tex_in;
- ^
-program_source:208:118: error: call to deleted constructor of 'texture2d<float, access::sample>'
- tint_module_vars_struct const tint_module_vars = tint_module_vars_struct{.ubo=ubo, .buf_out=buf_out, .tex_in=tex_in};
- ^
-/System/Library/PrivateFrameworks/GPUCompiler.framework/Versions/32023/Libraries/lib/clang/32023.194/include/metal/metal_texture:6083:3: note: 'texture2d' has been explicitly marked deleted here
- texture2d() thread = delete;
- ^
-program_source:51:36: note: in implicit initialization of field 'tint_symbol' with omitted initializer
- texture2d<float, access::sample> tint_symbol;
- ^
-program_source:213:136: error: call to deleted constructor of 'texture2d<float, access::sample>'
- tint_module_vars_struct const tint_module_vars = tint_module_vars_struct{.ubo=ubo, .buf_in=buf_in, .buf_out=buf_out, .tex_out=tex_out};
- ^
-/System/Library/PrivateFrameworks/GPUCompiler.framework/Versions/32023/Libraries/lib/clang/32023.194/include/metal/metal_texture:6083:3: note: 'texture2d' has been explicitly marked deleted here
- texture2d() thread = delete;
- ^
-program_source:51:36: note: in implicit initialization of field 'tint_symbol' with omitted initializer
- texture2d<float, access::sample> tint_symbol;
- ^
-
diff --git a/test/tint/bug/chromium/1434271.wgsl.expected.ir.msl b/test/tint/bug/chromium/1434271.wgsl.expected.ir.msl
index e15baf4..ac63654 100644
--- a/test/tint/bug/chromium/1434271.wgsl.expected.ir.msl
+++ b/test/tint/bug/chromium/1434271.wgsl.expected.ir.msl
@@ -1,5 +1,3 @@
-SKIP: FAILED
-
#include <metal_stdlib>
using namespace metal;
@@ -15,17 +13,6 @@
float2 quad_pos;
};
-struct RenderParams {
- float4x4 modelViewProjectionMatrix;
- float3 right;
- float3 up;
-};
-
-struct SimulationParams {
- float deltaTime;
- float4 seed;
-};
-
template<typename T, size_t N>
struct tint_array {
const constant T& operator[](size_t i) const constant { return elements[i]; }
@@ -38,30 +25,45 @@
T elements[N];
};
-struct Particle {
- float3 position;
- float lifetime;
- float4 color;
- float2 velocity;
+struct RenderParams_packed_vec3 {
+ /* 0x0000 */ float4x4 modelViewProjectionMatrix;
+ /* 0x0040 */ packed_float3 right;
+ /* 0x004c */ tint_array<int8_t, 4> tint_pad;
+ /* 0x0050 */ packed_float3 up;
+ /* 0x005c */ tint_array<int8_t, 4> tint_pad_1;
};
-struct Particles {
- tint_array<Particle, 1> particles;
+struct SimulationParams {
+ /* 0x0000 */ float deltaTime;
+ /* 0x0004 */ tint_array<int8_t, 12> tint_pad_2;
+ /* 0x0010 */ float4 seed;
+};
+
+struct Particle_packed_vec3 {
+ /* 0x0000 */ packed_float3 position;
+ /* 0x000c */ float lifetime;
+ /* 0x0010 */ float4 color;
+ /* 0x0020 */ float2 velocity;
+ /* 0x0028 */ tint_array<int8_t, 8> tint_pad_3;
+};
+
+struct Particles_packed_vec3 {
+ /* 0x0000 */ tint_array<Particle_packed_vec3, 1> particles;
};
struct UBO {
- uint width;
+ /* 0x0000 */ uint width;
};
struct Buffer {
- tint_array<float, 1> weights;
+ /* 0x0000 */ tint_array<float, 1> weights;
};
struct tint_module_vars_struct {
thread float2* rand_seed;
- const constant RenderParams* render_params;
+ const constant RenderParams_packed_vec3* render_params;
const constant SimulationParams* sim_params;
- device Particles* data;
+ device Particles_packed_vec3* data;
texture1d<float, access::sample> tint_symbol;
const constant UBO* ubo;
const device Buffer* buf_in;
@@ -70,6 +72,13 @@
texture2d<float, access::write> tex_out;
};
+struct Particle {
+ float3 position;
+ float lifetime;
+ float4 color;
+ float2 velocity;
+};
+
struct vertex_main_outputs {
float4 tint_symbol_1 [[position]];
};
@@ -105,35 +114,41 @@
}
VertexOutput vs_main_inner(VertexInput in, tint_module_vars_struct tint_module_vars) {
- float3 quad_pos = (float2x3((*tint_module_vars.render_params).right, (*tint_module_vars.render_params).up) * in.quad_pos);
+ float3 const v = float3((*tint_module_vars.render_params).right);
+ float3 quad_pos = (float2x3(v, float3((*tint_module_vars.render_params).up)) * in.quad_pos);
float3 position = (in.position - (quad_pos + 0.00999999977648258209f));
VertexOutput out = {};
- float4x4 const v = (*tint_module_vars.render_params).modelViewProjectionMatrix;
- out.position = (v * float4(position, 1.0f));
+ float4x4 const v_1 = (*tint_module_vars.render_params).modelViewProjectionMatrix;
+ out.position = (v_1 * float4(position, 1.0f));
out.color = in.color;
out.quad_pos = in.quad_pos;
return out;
}
-void tint_store_and_preserve_padding(device Particle* const target, Particle value_param) {
- (*target).position = value_param.position;
+void tint_store_and_preserve_padding(device Particle_packed_vec3* const target, Particle value_param) {
+ (*target).position = packed_float3(value_param.position);
(*target).lifetime = value_param.lifetime;
(*target).color = value_param.color;
(*target).velocity = value_param.velocity;
}
+Particle tint_load_struct_packed_vec3(device Particle_packed_vec3* const from) {
+ float3 const v_2 = float3((*from).position);
+ return Particle{.position=v_2, .lifetime=(*from).lifetime, .color=(*from).color, .velocity=(*from).velocity};
+}
+
void simulate_inner(uint3 GlobalInvocationID, tint_module_vars_struct tint_module_vars) {
- float2 const v_1 = (*tint_module_vars.sim_params).seed.xy;
- float2 const v_2 = (v_1 * float2(GlobalInvocationID.xy));
- (*tint_module_vars.rand_seed) = (v_2 * (*tint_module_vars.sim_params).seed.zw);
+ float2 const v_3 = (*tint_module_vars.sim_params).seed.xy;
+ float2 const v_4 = (v_3 * float2(GlobalInvocationID.xy));
+ (*tint_module_vars.rand_seed) = (v_4 * (*tint_module_vars.sim_params).seed.zw);
uint const idx = GlobalInvocationID[0u];
- Particle particle = (*tint_module_vars.data).particles[idx];
+ Particle particle = tint_load_struct_packed_vec3((&(*tint_module_vars.data).particles[idx]));
tint_store_and_preserve_padding((&(*tint_module_vars.data).particles[idx]), particle);
}
void export_level_inner(uint3 coord, tint_module_vars_struct tint_module_vars) {
- uint const v_3 = tint_module_vars.tex_out.get_width(0u);
- if (all((coord.xy < uint2(uint2(v_3, tint_module_vars.tex_out.get_height(0u)))))) {
+ uint const v_5 = tint_module_vars.tex_out.get_width(0u);
+ if (all((coord.xy < uint2(uint2(v_5, tint_module_vars.tex_out.get_height(0u)))))) {
uint const dst_offset = (coord[0u] << ((coord[1u] * (*tint_module_vars.ubo).width) & 31u));
uint const src_offset = ((coord[0u] - 2u) + ((coord[1u] >> (2u & 31u)) * (*tint_module_vars.ubo).width));
float const a = (*tint_module_vars.buf_in).weights[(src_offset << (0u & 31u))];
@@ -142,23 +157,29 @@
float const d = (*tint_module_vars.buf_in).weights[((src_offset + 1u) + (*tint_module_vars.ubo).width)];
float const sum = dot(float4(a, b, c, d), float4(1.0f));
(*tint_module_vars.buf_out).weights[dst_offset] = fmod(sum, 4.0f);
- float4 const v_4 = float4(a, (a * b), ((a / b) + c), sum);
- float4 const probabilities = (v_4 + max(sum, 0.0f));
+ float4 const v_6 = float4(a, (a * b), ((a / b) + c), sum);
+ float4 const probabilities = (v_6 + max(sum, 0.0f));
tint_module_vars.tex_out.write(probabilities, uint2(int2(coord.xy)));
}
}
vertex vertex_main_outputs vertex_main() {
- return vertex_main_outputs{.tint_symbol_1=vertex_main_inner()};
+ vertex_main_outputs tint_wrapper_result = {};
+ tint_wrapper_result.tint_symbol_1 = vertex_main_inner();
+ return tint_wrapper_result;
}
-vertex vs_main_outputs vs_main(vs_main_inputs inputs [[stage_in]], const constant RenderParams* render_params [[buffer(0)]]) {
+vertex vs_main_outputs vs_main(vs_main_inputs inputs [[stage_in]], const constant RenderParams_packed_vec3* render_params [[buffer(0)]]) {
tint_module_vars_struct const tint_module_vars = tint_module_vars_struct{.render_params=render_params};
- VertexOutput const v_5 = vs_main_inner(VertexInput{.position=inputs.VertexInput_position, .color=inputs.VertexInput_color, .quad_pos=inputs.VertexInput_quad_pos}, tint_module_vars);
- return vs_main_outputs{.VertexOutput_position=v_5.position, .VertexOutput_color=v_5.color, .VertexOutput_quad_pos=v_5.quad_pos};
+ VertexOutput const v_7 = vs_main_inner(VertexInput{.position=inputs.VertexInput_position, .color=inputs.VertexInput_color, .quad_pos=inputs.VertexInput_quad_pos}, tint_module_vars);
+ vs_main_outputs tint_wrapper_result = {};
+ tint_wrapper_result.VertexOutput_position = v_7.position;
+ tint_wrapper_result.VertexOutput_color = v_7.color;
+ tint_wrapper_result.VertexOutput_quad_pos = v_7.quad_pos;
+ return tint_wrapper_result;
}
-kernel void simulate(uint3 GlobalInvocationID [[thread_position_in_grid]], const constant SimulationParams* sim_params [[buffer(1)]], device Particles* data [[buffer(2)]]) {
+kernel void simulate(uint3 GlobalInvocationID [[thread_position_in_grid]], const constant SimulationParams* sim_params [[buffer(1)]], device Particles_packed_vec3* data [[buffer(2)]]) {
thread float2 rand_seed = 0.0f;
tint_module_vars_struct const tint_module_vars = tint_module_vars_struct{.rand_seed=(&rand_seed), .sim_params=sim_params, .data=data};
simulate_inner(GlobalInvocationID, tint_module_vars);
@@ -168,34 +189,3 @@
tint_module_vars_struct const tint_module_vars = tint_module_vars_struct{.ubo=ubo, .buf_in=buf_in, .buf_out=buf_out, .tex_out=tex_out};
export_level_inner(coord, tint_module_vars);
}
-program_source:89:8: warning: unused variable 'res' [-Wunused-variable]
- half res = asinh(arg_0);
- ^
-program_source:154:104: error: call to deleted constructor of 'texture1d<float, access::sample>'
- tint_module_vars_struct const tint_module_vars = tint_module_vars_struct{.render_params=render_params};
- ^
-/System/Library/PrivateFrameworks/GPUCompiler.framework/Versions/32023/Libraries/lib/clang/32023.194/include/metal/metal_texture:2308:3: note: 'texture1d' has been explicitly marked deleted here
- texture1d() thread = delete;
- ^
-program_source:63:36: note: in implicit initialization of field 'tint_symbol' with omitted initializer
- texture1d<float, access::sample> tint_symbol;
- ^
-program_source:161:135: error: call to deleted constructor of 'texture1d<float, access::sample>'
- tint_module_vars_struct const tint_module_vars = tint_module_vars_struct{.rand_seed=(&rand_seed), .sim_params=sim_params, .data=data};
- ^
-/System/Library/PrivateFrameworks/GPUCompiler.framework/Versions/32023/Libraries/lib/clang/32023.194/include/metal/metal_texture:2308:3: note: 'texture1d' has been explicitly marked deleted here
- texture1d() thread = delete;
- ^
-program_source:63:36: note: in implicit initialization of field 'tint_symbol' with omitted initializer
- texture1d<float, access::sample> tint_symbol;
- ^
-program_source:166:136: error: call to deleted constructor of 'texture1d<float, access::sample>'
- tint_module_vars_struct const tint_module_vars = tint_module_vars_struct{.ubo=ubo, .buf_in=buf_in, .buf_out=buf_out, .tex_out=tex_out};
- ^
-/System/Library/PrivateFrameworks/GPUCompiler.framework/Versions/32023/Libraries/lib/clang/32023.194/include/metal/metal_texture:2308:3: note: 'texture1d' has been explicitly marked deleted here
- texture1d() thread = delete;
- ^
-program_source:63:36: note: in implicit initialization of field 'tint_symbol' with omitted initializer
- texture1d<float, access::sample> tint_symbol;
- ^
-
diff --git a/test/tint/bug/dawn/947.wgsl.expected.ir.msl b/test/tint/bug/dawn/947.wgsl.expected.ir.msl
index 40d3a71..a6ccc01 100644
--- a/test/tint/bug/dawn/947.wgsl.expected.ir.msl
+++ b/test/tint/bug/dawn/947.wgsl.expected.ir.msl
@@ -1,5 +1,3 @@
-SKIP: FAILED
-
#include <metal_stdlib>
using namespace metal;
@@ -9,8 +7,8 @@
};
struct Uniforms {
- float2 u_scale;
- float2 u_offset;
+ /* 0x0000 */ float2 u_scale;
+ /* 0x0008 */ float2 u_offset;
};
struct tint_module_vars_struct {
@@ -73,30 +71,16 @@
vertex vs_main_outputs vs_main(uint VertexIndex [[vertex_id]], const constant Uniforms* uniforms [[buffer(0)]]) {
tint_module_vars_struct const tint_module_vars = tint_module_vars_struct{.uniforms=uniforms};
VertexOutputs const v = vs_main_inner(VertexIndex, tint_module_vars);
- return vs_main_outputs{.VertexOutputs_texcoords=v.texcoords, .VertexOutputs_position=v.position};
+ vs_main_outputs tint_wrapper_result = {};
+ tint_wrapper_result.VertexOutputs_texcoords = v.texcoords;
+ tint_wrapper_result.VertexOutputs_position = v.position;
+ return tint_wrapper_result;
}
fragment fs_main_outputs fs_main(fs_main_inputs inputs [[stage_in]]) {
thread bool continue_execution = true;
tint_module_vars_struct const tint_module_vars = tint_module_vars_struct{.continue_execution=(&continue_execution)};
- return fs_main_outputs{.tint_symbol=fs_main_inner(inputs.texcoord, tint_module_vars)};
+ fs_main_outputs tint_wrapper_result = {};
+ tint_wrapper_result.tint_symbol = fs_main_inner(inputs.texcoord, tint_module_vars);
+ return tint_wrapper_result;
}
-program_source:72:94: error: call to deleted constructor of 'texture2d<float, access::sample>'
- tint_module_vars_struct const tint_module_vars = tint_module_vars_struct{.uniforms=uniforms};
- ^
-/System/Library/PrivateFrameworks/GPUCompiler.framework/Versions/32023/Libraries/lib/clang/32023.194/include/metal/metal_texture:6083:3: note: 'texture2d' has been explicitly marked deleted here
- texture2d() thread = delete;
- ^
-program_source:17:36: note: in implicit initialization of field 'myTexture' with omitted initializer
- texture2d<float, access::sample> myTexture;
- ^
-program_source:79:117: error: call to deleted constructor of 'texture2d<float, access::sample>'
- tint_module_vars_struct const tint_module_vars = tint_module_vars_struct{.continue_execution=(&continue_execution)};
- ^
-/System/Library/PrivateFrameworks/GPUCompiler.framework/Versions/32023/Libraries/lib/clang/32023.194/include/metal/metal_texture:6083:3: note: 'texture2d' has been explicitly marked deleted here
- texture2d() thread = delete;
- ^
-program_source:17:36: note: in implicit initialization of field 'myTexture' with omitted initializer
- texture2d<float, access::sample> myTexture;
- ^
-
diff --git a/test/tint/bug/fxc/gradient_in_varying_loop/1112.wgsl.expected.ir.msl b/test/tint/bug/fxc/gradient_in_varying_loop/1112.wgsl.expected.ir.msl
index 55a628e..4868c1c 100644
--- a/test/tint/bug/fxc/gradient_in_varying_loop/1112.wgsl.expected.ir.msl
+++ b/test/tint/bug/fxc/gradient_in_varying_loop/1112.wgsl.expected.ir.msl
@@ -1,5 +1,3 @@
-SKIP: FAILED
-
#include <metal_stdlib>
using namespace metal;
@@ -9,6 +7,9 @@
texture2d<float, access::sample> depthTexture;
};
+#define TINT_ISOLATE_UB(VOLATILE_NAME) \
+ {volatile bool VOLATILE_NAME = false; if (VOLATILE_NAME) break;}
+
struct tint_symbol_outputs {
float4 tint_symbol_1 [[color(0)]];
};
@@ -22,6 +23,7 @@
int i = 0;
{
while(true) {
+ TINT_ISOLATE_UB(tint_volatile_false)
if ((i < 1)) {
} else {
break;
@@ -46,13 +48,13 @@
v_2 = (offset[1u] > 1.0f);
}
if (v_2) {
- i = (i + 1);
+ i = as_type<int>((as_type<uint>(i) + as_type<uint>(1)));
{
}
continue;
}
float const sampleDepth = 0.0f;
- i = (i + 1);
+ i = as_type<int>((as_type<uint>(i) + as_type<uint>(1)));
{
}
continue;
@@ -63,18 +65,7 @@
fragment tint_symbol_outputs tint_symbol(tint_symbol_inputs inputs [[stage_in]], sampler Sampler [[sampler(0)]], texture2d<float, access::sample> randomTexture [[texture(0)]]) {
tint_module_vars_struct const tint_module_vars = tint_module_vars_struct{.Sampler=Sampler, .randomTexture=randomTexture};
- return tint_symbol_outputs{.tint_symbol_1=tint_symbol_inner(inputs.vUV, tint_module_vars)};
+ tint_symbol_outputs tint_wrapper_result = {};
+ tint_wrapper_result.tint_symbol_1 = tint_symbol_inner(inputs.vUV, tint_module_vars);
+ return tint_wrapper_result;
}
-program_source:52:19: warning: unused variable 'sampleDepth' [-Wunused-variable]
- float const sampleDepth = 0.0f;
- ^
-program_source:63:122: error: call to deleted constructor of 'texture2d<float, access::sample>'
- tint_module_vars_struct const tint_module_vars = tint_module_vars_struct{.Sampler=Sampler, .randomTexture=randomTexture};
- ^
-/System/Library/PrivateFrameworks/GPUCompiler.framework/Versions/32023/Libraries/lib/clang/32023.194/include/metal/metal_texture:6083:3: note: 'texture2d' has been explicitly marked deleted here
- texture2d() thread = delete;
- ^
-program_source:7:36: note: in implicit initialization of field 'depthTexture' with omitted initializer
- texture2d<float, access::sample> depthTexture;
- ^
-
diff --git a/test/tint/bug/tint/949.wgsl.expected.ir.msl b/test/tint/bug/tint/949.wgsl.expected.ir.msl
index 9c6169c..10f6032 100644
--- a/test/tint/bug/tint/949.wgsl.expected.ir.msl
+++ b/test/tint/bug/tint/949.wgsl.expected.ir.msl
@@ -1,5 +1,3 @@
-SKIP: FAILED
-
#include <metal_stdlib>
using namespace metal;
@@ -8,26 +6,40 @@
float3 specular;
};
-struct LeftOver {
- float4x4 u_World;
- float4x4 u_ViewProjection;
- float u_bumpStrength;
- uint padding;
- float3 u_cameraPosition;
- float u_parallaxScale;
- float textureInfoName;
- uint padding_1;
- float2 tangentSpaceParameter0;
+template<typename T, size_t N>
+struct tint_array {
+ const constant T& operator[](size_t i) const constant { return elements[i]; }
+ device T& operator[](size_t i) device { return elements[i]; }
+ const device T& operator[](size_t i) const device { return elements[i]; }
+ thread T& operator[](size_t i) thread { return elements[i]; }
+ const thread T& operator[](size_t i) const thread { return elements[i]; }
+ threadgroup T& operator[](size_t i) threadgroup { return elements[i]; }
+ const threadgroup T& operator[](size_t i) const threadgroup { return elements[i]; }
+ T elements[N];
};
-struct Light0 {
- float4 vLightData;
- float4 vLightDiffuse;
- float4 vLightSpecular;
- float3 vLightGround;
- uint padding_2;
- float4 shadowsInfo;
- float2 depthValues;
+struct LeftOver_packed_vec3 {
+ /* 0x0000 */ float4x4 u_World;
+ /* 0x0040 */ float4x4 u_ViewProjection;
+ /* 0x0080 */ float u_bumpStrength;
+ /* 0x0084 */ uint padding;
+ /* 0x0088 */ tint_array<int8_t, 8> tint_pad;
+ /* 0x0090 */ packed_float3 u_cameraPosition;
+ /* 0x009c */ float u_parallaxScale;
+ /* 0x00a0 */ float textureInfoName;
+ /* 0x00a4 */ uint padding_1;
+ /* 0x00a8 */ float2 tangentSpaceParameter0;
+};
+
+struct Light0_packed_vec3 {
+ /* 0x0000 */ float4 vLightData;
+ /* 0x0010 */ float4 vLightDiffuse;
+ /* 0x0020 */ float4 vLightSpecular;
+ /* 0x0030 */ packed_float3 vLightGround;
+ /* 0x003c */ uint padding_2;
+ /* 0x0040 */ float4 shadowsInfo;
+ /* 0x0050 */ float2 depthValues;
+ /* 0x0058 */ tint_array<int8_t, 8> tint_pad_1;
};
struct tint_module_vars_struct {
@@ -36,19 +48,22 @@
texture2d<float, access::sample> TextureSamplerTexture;
sampler TextureSamplerSampler;
thread float2* vMainuv;
- const constant LeftOver* x_269;
+ const constant LeftOver_packed_vec3* x_269;
thread float4* v_output1;
thread bool* gl_FrontFacing;
thread float2* v_uv;
thread float4* v_output2;
texture2d<float, access::sample> TextureSampler1Texture;
sampler TextureSampler1Sampler;
- const constant Light0* light0;
+ const constant Light0_packed_vec3* light0;
thread float4* glFragColor;
sampler bumpSamplerSampler;
texture2d<float, access::sample> bumpSamplerTexture;
};
+#define TINT_ISOLATE_UB(VOLATILE_NAME) \
+ {volatile bool VOLATILE_NAME = false; if (VOLATILE_NAME) break;}
+
struct main_out {
float4 glFragColor_1;
};
@@ -262,7 +277,7 @@
float4 const x_264 = tempTextureRead;
float const x_273 = (*tint_module_vars.x_269).textureInfoName;
rgb = (float3(x_264[0u], x_264[1u], x_264[2u]) * x_273);
- float3 const x_279 = (*tint_module_vars.x_269).u_cameraPosition;
+ float3 const x_279 = float3((*tint_module_vars.x_269).u_cameraPosition);
float4 const x_282 = (*tint_module_vars.v_output1);
output5 = normalize((x_279 - float3(x_282[0u], x_282[1u], x_282[2u])));
output4 = float4(0.0f);
@@ -325,6 +340,7 @@
i = 0;
{
while(true) {
+ TINT_ISOLATE_UB(tint_volatile_false)
int const x_388 = i;
if ((x_388 < 15)) {
} else {
@@ -369,7 +385,7 @@
}
{
int const x_441 = i;
- i = (x_441 + 1);
+ i = as_type<int>((as_type<uint>(x_441) + as_type<uint>(1)));
}
continue;
}
@@ -397,7 +413,7 @@
tempTextureRead1 = x_475;
float4 const x_477 = tempTextureRead1;
rgb1 = float3(x_477[0u], x_477[1u], x_477[2u]);
- float3 const x_481 = (*tint_module_vars.x_269).u_cameraPosition;
+ float3 const x_481 = float3((*tint_module_vars.x_269).u_cameraPosition);
float4 const x_482 = (*tint_module_vars.v_output1);
viewDirectionW_1 = normalize((x_481 - float3(x_482[0u], x_482[1u], x_482[2u])));
shadow = 1.0f;
@@ -417,7 +433,7 @@
param_14 = float3(x_510[0u], x_510[1u], x_510[2u]);
float4 const x_514 = (*tint_module_vars.light0).vLightSpecular;
param_15 = float3(x_514[0u], x_514[1u], x_514[2u]);
- float3 const x_518 = (*tint_module_vars.light0).vLightGround;
+ float3 const x_518 = float3((*tint_module_vars.light0).vLightGround);
param_16 = x_518;
float const x_520 = glossiness_1;
param_17 = x_520;
@@ -455,7 +471,7 @@
return main_out{.glFragColor_1=(*tint_module_vars.glFragColor)};
}
-fragment tint_symbol_outputs tint_symbol(tint_symbol_inputs inputs [[stage_in]], bool gl_FrontFacing_param [[front_facing]], texture2d<float, access::sample> TextureSamplerTexture [[texture(0)]], sampler TextureSamplerSampler [[sampler(0)]], const constant LeftOver* x_269 [[buffer(0)]], texture2d<float, access::sample> TextureSampler1Texture [[texture(1)]], sampler TextureSampler1Sampler [[sampler(1)]], const constant Light0* light0 [[buffer(1)]]) {
+fragment tint_symbol_outputs tint_symbol(tint_symbol_inputs inputs [[stage_in]], bool gl_FrontFacing_param [[front_facing]], texture2d<float, access::sample> TextureSamplerTexture [[texture(0)]], sampler TextureSamplerSampler [[sampler(0)]], const constant LeftOver_packed_vec3* x_269 [[buffer(0)]], texture2d<float, access::sample> TextureSampler1Texture [[texture(1)]], sampler TextureSampler1Sampler [[sampler(1)]], const constant Light0_packed_vec3* light0 [[buffer(1)]]) {
thread float u_Float = 0.0f;
thread float3 u_Color = 0.0f;
thread float2 vMainuv = 0.0f;
@@ -465,21 +481,7 @@
thread float4 v_output2 = 0.0f;
thread float4 glFragColor = 0.0f;
tint_module_vars_struct const tint_module_vars = tint_module_vars_struct{.u_Float=(&u_Float), .u_Color=(&u_Color), .TextureSamplerTexture=TextureSamplerTexture, .TextureSamplerSampler=TextureSamplerSampler, .vMainuv=(&vMainuv), .x_269=x_269, .v_output1=(&v_output1), .gl_FrontFacing=(&gl_FrontFacing), .v_uv=(&v_uv), .v_output2=(&v_output2), .TextureSampler1Texture=TextureSampler1Texture, .TextureSampler1Sampler=TextureSampler1Sampler, .light0=light0, .glFragColor=(&glFragColor)};
- return tint_symbol_outputs{.main_out_glFragColor_1=tint_symbol_inner(inputs.vMainuv_param, inputs.v_output1_param, gl_FrontFacing_param, inputs.v_uv_param, inputs.v_output2_param, tint_module_vars).glFragColor_1};
+ tint_symbol_outputs tint_wrapper_result = {};
+ tint_wrapper_result.main_out_glFragColor_1 = tint_symbol_inner(inputs.vMainuv_param, inputs.v_output1_param, gl_FrontFacing_param, inputs.v_uv_param, inputs.v_output2_param, tint_module_vars).glFragColor_1;
+ return tint_wrapper_result;
}
-program_source:331:20: warning: unused variable 'x_394' [-Wunused-variable]
- float2 const x_394 = (*tint_module_vars.v_uv);
- ^
-program_source:332:20: warning: unused variable 'x_395' [-Wunused-variable]
- float2 const x_395 = vCurrOffset;
- ^
-program_source:465:484: error: call to deleted constructor of 'texture2d<float, access::sample>'
- tint_module_vars_struct const tint_module_vars = tint_module_vars_struct{.u_Float=(&u_Float), .u_Color=(&u_Color), .TextureSamplerTexture=TextureSamplerTexture, .TextureSamplerSampler=TextureSamplerSampler, .vMainuv=(&vMainuv), .x_269=x_269, .v_output1=(&v_output1), .gl_FrontFacing=(&gl_FrontFacing), .v_uv=(&v_uv), .v_output2=(&v_output2), .TextureSampler1Texture=TextureSampler1Texture, .TextureSampler1Sampler=TextureSampler1Sampler, .light0=light0, .glFragColor=(&glFragColor)};
- ^
-/System/Library/PrivateFrameworks/GPUCompiler.framework/Versions/32023/Libraries/lib/clang/32023.194/include/metal/metal_texture:6083:3: note: 'texture2d' has been explicitly marked deleted here
- texture2d() thread = delete;
- ^
-program_source:47:36: note: in implicit initialization of field 'bumpSamplerTexture' with omitted initializer
- texture2d<float, access::sample> bumpSamplerTexture;
- ^
-
