blob: 0e29de7b363c8a4bb5cad716561974c20e522e08 [file] [log] [blame] [edit]
#include <metal_stdlib>
using namespace metal;
void tint_zero_workgroup_memory(uint local_idx, threadgroup atomic_int* const tint_symbol_11) {
if ((local_idx < 1u)) {
atomic_store_explicit(tint_symbol_11, 0, memory_order_relaxed);
}
threadgroup_barrier(mem_flags::mem_threadgroup);
}
int tint_ftoi(float v) {
return select(2147483647, select(int(v), (-2147483647 - 1), (v < -2147483648.0f)), (v < 2147483520.0f));
}
struct S {
int a;
int b;
};
int accept_value(int val) {
return val;
}
int accept_ptr_deref_call_func(thread int* const val) {
int const tint_symbol_3 = *(val);
int const tint_symbol_4 = accept_value(*(val));
return as_type<int>((as_type<uint>(tint_symbol_3) + as_type<uint>(tint_symbol_4)));
}
int accept_ptr_deref_pass_through(thread int* const val) {
int const tint_symbol_1 = *(val);
int const tint_symbol_2 = accept_ptr_deref_call_func(val);
return as_type<int>((as_type<uint>(tint_symbol_1) + as_type<uint>(tint_symbol_2)));
}
int accept_ptr_to_struct_and_access(thread S* const val) {
return as_type<int>((as_type<uint>((*(val)).a) + as_type<uint>((*(val)).b)));
}
int accept_ptr_to_struct_access_pass_ptr(thread S* const val) {
(*(val)).a = 2;
return (*(val)).a;
}
int accept_ptr_vec_access_elements(thread float3* const v1) {
(*(v1))[0] = cross(*(v1), *(v1))[0];
return tint_ftoi((*(v1))[0]);
}
int call_builtin_with_mod_scope_ptr(threadgroup atomic_int* const tint_symbol_12) {
return atomic_load_explicit(tint_symbol_12, memory_order_relaxed);
}
void tint_symbol_inner(uint local_invocation_index, threadgroup atomic_int* const tint_symbol_13, device int* const tint_symbol_14) {
tint_zero_workgroup_memory(local_invocation_index, tint_symbol_13);
int v1 = 0;
S v2 = S{};
float3 v4 = float3(0.0f);
int const t1 = atomic_load_explicit(tint_symbol_13, memory_order_relaxed);
int const tint_symbol_5 = accept_ptr_deref_pass_through(&(v1));
int const tint_symbol_6 = accept_ptr_to_struct_and_access(&(v2));
int const tint_symbol_7 = accept_ptr_to_struct_and_access(&(v2));
int const tint_symbol_8 = accept_ptr_vec_access_elements(&(v4));
int const tint_symbol_9 = accept_ptr_to_struct_access_pass_ptr(&(v2));
int const tint_symbol_10 = call_builtin_with_mod_scope_ptr(tint_symbol_13);
*(tint_symbol_14) = as_type<int>((as_type<uint>(as_type<int>((as_type<uint>(as_type<int>((as_type<uint>(as_type<int>((as_type<uint>(as_type<int>((as_type<uint>(as_type<int>((as_type<uint>(tint_symbol_5) + as_type<uint>(tint_symbol_6)))) + as_type<uint>(tint_symbol_7)))) + as_type<uint>(tint_symbol_8)))) + as_type<uint>(tint_symbol_9)))) + as_type<uint>(tint_symbol_10)))) + as_type<uint>(t1)));
}
kernel void tint_symbol(device int* tint_symbol_16 [[buffer(0)]], uint local_invocation_index [[thread_index_in_threadgroup]]) {
threadgroup atomic_int tint_symbol_15;
tint_symbol_inner(local_invocation_index, &(tint_symbol_15), tint_symbol_16);
return;
}