test/tint/bug/chromium/1434271.wgsl.expected.ir.msl - dawn - Git at Google

 SKIP: FAILED

 #include <metal_stdlib>
 using namespace metal;

 struct VertexOutput {
   float4 position;
   float4 color;
   float2 quad_pos;
 };

 struct VertexInput {
   float3 position;
   float4 color;
   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]; }
   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;
   float2 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;
   texture1d<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 vertex_main_outputs {
   float4 tint_symbol_1 [[position]];
 };

 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)]];
 };

 void asinh_468a48() {
   half arg_0 = 0.0h;
   half res = asinh(arg_0);
 }

 float4 vertex_main_inner() {
   asinh_468a48();
   return float4(0.0f);
 }

 fragment void fragment_main() {
   asinh_468a48();
 }

 kernel void rgba32uintin() {
   asinh_468a48();
 }

 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;
 }

 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_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);
   uint const idx = GlobalInvocationID[0u];
   Particle particle = (*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 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))];
     float const b = (*tint_module_vars.buf_in).weights[(src_offset + 1u)];
     float const c = (*tint_module_vars.buf_in).weights[((src_offset + 1u) + (*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] = 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));
     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 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_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};
 }

 kernel void simulate(uint3 GlobalInvocationID [[thread_position_in_grid]], const constant SimulationParams* sim_params [[buffer(1)]], device Particles* 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);
 }

 kernel void export_level(uint3 coord [[thread_position_in_grid]], const constant UBO* ubo [[buffer(3)]], const device Buffer* buf_in [[buffer(4)]], device Buffer* buf_out [[buffer(0)]], texture2d<float, access::write> tex_out [[texture(0)]]) {
   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;
                                    ^
	SKIP: FAILED

	#include <metal_stdlib>
	using namespace metal;

	struct VertexOutput {
	float4 position;
	float4 color;
	float2 quad_pos;
	};

	struct VertexInput {
	float3 position;
	float4 color;
	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]; }
	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;
	float2 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;
	texture1d<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 vertex_main_outputs {
	float4 tint_symbol_1 [[position]];
	};

	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)]];
	};

	void asinh_468a48() {
	half arg_0 = 0.0h;
	half res = asinh(arg_0);
	}

	float4 vertex_main_inner() {
	asinh_468a48();
	return float4(0.0f);
	}

	fragment void fragment_main() {
	asinh_468a48();
	}

	kernel void rgba32uintin() {
	asinh_468a48();
	}

	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;
	}

	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_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);
	uint const idx = GlobalInvocationID[0u];
	Particle particle = (*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 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))];
	float const b = (*tint_module_vars.buf_in).weights[(src_offset + 1u)];
	float const c = (tint_module_vars.buf_in).weights[((src_offset + 1u) + (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] = 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));
	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 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_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};
	}

	kernel void simulate(uint3 GlobalInvocationID [[thread_position_in_grid]], const constant SimulationParams* sim_params [[buffer(1)]], device Particles* 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);
	}

	kernel void export_level(uint3 coord [[thread_position_in_grid]], const constant UBO* ubo [[buffer(3)]], const device Buffer* buf_in [[buffer(4)]], device Buffer* buf_out [[buffer(0)]], texture2d<float, access::write> tex_out [[texture(0)]]) {
	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;
	^