dawn-cmake/test/benchmark/particles.wgsl.expected.msl

#include <metal_stdlib>

using namespace metal;

template<typename T, int N, int M>
inline vec<T, M> operator*(matrix<T, N, M> lhs, packed_vec<T, N> rhs) {
  return lhs * vec<T, N>(rhs);
}

template<typename T, int N, int M>
inline vec<T, N> operator*(packed_vec<T, M> lhs, matrix<T, N, M> rhs) {
  return vec<T, M>(lhs) * rhs;
}

struct RenderParams {
  /* 0x0000 */ float4x4 modelViewProjectionMatrix;
  /* 0x0040 */ packed_float3 right;
  /* 0x004c */ int8_t tint_pad[4];
  /* 0x0050 */ packed_float3 up;
  /* 0x005c */ int8_t tint_pad_1[4];
};
struct VertexInput {
  float3 position;
  float4 color;
  float2 quad_pos;
};
struct VertexOutput {
  float4 position;
  float4 color;
  float2 quad_pos;
};
struct tint_symbol_2 {
  float3 position [[attribute(0)]];
  float4 color [[attribute(1)]];
  float2 quad_pos [[attribute(2)]];
};
struct tint_symbol_3 {
  float4 color [[user(locn0)]];
  float2 quad_pos [[user(locn1)]];
  float4 position [[position]];
};
struct tint_symbol_5 {
  float4 color [[user(locn0)]];
  float2 quad_pos [[user(locn1)]];
};
struct tint_symbol_6 {
  float4 value [[color(0)]];
};
struct SimulationParams {
  /* 0x0000 */ float deltaTime;
  /* 0x0004 */ int8_t tint_pad_2[12];
  /* 0x0010 */ float4 seed;
};
struct Particle {
  /* 0x0000 */ packed_float3 position;
  /* 0x000c */ float lifetime;
  /* 0x0010 */ float4 color;
  /* 0x0020 */ packed_float3 velocity;
  /* 0x002c */ int8_t tint_pad_3[4];
};
struct Particles {
  /* 0x0000 */ Particle particles[1];
};
struct UBO {
  /* 0x0000 */ uint width;
};
struct Buffer {
  /* 0x0000 */ float weights[1];
};

float rand(thread float2* const tint_symbol_9) {
  (*(tint_symbol_9))[0] = fract((cos(dot(*(tint_symbol_9), float2(23.140779495f, 232.616897583f))) * 136.816802979f));
  (*(tint_symbol_9))[1] = fract((cos(dot(*(tint_symbol_9), float2(54.478565216f, 345.841522217f))) * 534.764526367f));
  return (*(tint_symbol_9))[1];
}

VertexOutput vs_main_inner(VertexInput in, const constant RenderParams* const tint_symbol_10) {
  float3 quad_pos = (float2x3((*(tint_symbol_10)).right, (*(tint_symbol_10)).up) * in.quad_pos);
  float3 position = (in.position + (quad_pos * 0.01f));
  VertexOutput out = {};
  out.position = ((*(tint_symbol_10)).modelViewProjectionMatrix * float4(position, 1.0f));
  out.color = in.color;
  out.quad_pos = in.quad_pos;
  return out;
}

vertex tint_symbol_3 vs_main(const constant RenderParams* tint_symbol_11 [[buffer(0)]], tint_symbol_2 tint_symbol_1 [[stage_in]]) {
  VertexInput const tint_symbol_7 = {.position=tint_symbol_1.position, .color=tint_symbol_1.color, .quad_pos=tint_symbol_1.quad_pos};
  VertexOutput const inner_result = vs_main_inner(tint_symbol_7, tint_symbol_11);
  tint_symbol_3 wrapper_result = {};
  wrapper_result.position = inner_result.position;
  wrapper_result.color = inner_result.color;
  wrapper_result.quad_pos = inner_result.quad_pos;
  return wrapper_result;
}

float4 fs_main_inner(VertexOutput in) {
  float4 color = in.color;
  color[3] = (color[3] * fmax((1.0f - length(in.quad_pos)), 0.0f));
  return color;
}

fragment tint_symbol_6 fs_main(float4 position [[position]], tint_symbol_5 tint_symbol_4 [[stage_in]]) {
  VertexOutput const tint_symbol_8 = {.position=position, .color=tint_symbol_4.color, .quad_pos=tint_symbol_4.quad_pos};
  float4 const inner_result_1 = fs_main_inner(tint_symbol_8);
  tint_symbol_6 wrapper_result_1 = {};
  wrapper_result_1.value = inner_result_1;
  return wrapper_result_1;
}

void simulate_inner(uint3 GlobalInvocationID, thread float2* const tint_symbol_12, const constant SimulationParams* const tint_symbol_13, device Particles* const tint_symbol_14, texture2d<float, access::sample> tint_symbol_15) {
  *(tint_symbol_12) = ((float4((*(tint_symbol_13)).seed).xy + float2(uint3(GlobalInvocationID).xy)) * float4((*(tint_symbol_13)).seed).zw);
  uint const idx = GlobalInvocationID[0];
  Particle particle = (*(tint_symbol_14)).particles[idx];
  particle.velocity[2] = (particle.velocity[2] - ((*(tint_symbol_13)).deltaTime * 0.5f));
  particle.position = (particle.position + ((*(tint_symbol_13)).deltaTime * particle.velocity));
  particle.lifetime = (particle.lifetime - (*(tint_symbol_13)).deltaTime);
  particle.color[3] = smoothstep(0.0f, 0.5f, particle.lifetime);
  if ((particle.lifetime < 0.0f)) {
    int2 coord = int2(0, 0);
    for(int level = as_type<int>((as_type<uint>(int(tint_symbol_15.get_num_mip_levels())) - as_type<uint>(1))); (level > 0); level = as_type<int>((as_type<uint>(level) - as_type<uint>(1)))) {
      float4 const probabilites = tint_symbol_15.read(uint2(coord), level);
      float4 const value = float4(rand(tint_symbol_12));
      bool4 const mask = ((value >= float4(0.0f, float4(probabilites).xyz)) & (value < probabilites));
      coord = as_type<int2>((as_type<uint2>(coord) * as_type<uint>(2)));
      coord[0] = as_type<int>((as_type<uint>(coord[0]) + as_type<uint>(select(0, 1, any(bool4(mask).yw)))));
      coord[1] = as_type<int>((as_type<uint>(coord[1]) + as_type<uint>(select(0, 1, any(bool4(mask).zw)))));
    }
    float2 const uv = (float2(coord) / float2(int2(tint_symbol_15.get_width(), tint_symbol_15.get_height())));
    particle.position = float3((((uv - 0.5f) * 3.0f) * float2(1.0f, -1.0f)), 0.0f);
    particle.color = tint_symbol_15.read(uint2(coord), 0);
    particle.velocity[0] = ((rand(tint_symbol_12) - 0.5f) * 0.100000001f);
    particle.velocity[1] = ((rand(tint_symbol_12) - 0.5f) * 0.100000001f);
    particle.velocity[2] = (rand(tint_symbol_12) * 0.300000012f);
    particle.lifetime = (0.5f + (rand(tint_symbol_12) * 2.0f));
  }
  (*(tint_symbol_14)).particles[idx] = particle;
}

kernel void simulate(const constant SimulationParams* tint_symbol_17 [[buffer(0)]], device Particles* tint_symbol_18 [[buffer(1)]], texture2d<float, access::sample> tint_symbol_19 [[texture(0)]], uint3 GlobalInvocationID [[thread_position_in_grid]]) {
  thread float2 tint_symbol_16 = 0.0f;
  simulate_inner(GlobalInvocationID, &(tint_symbol_16), tint_symbol_17, tint_symbol_18, tint_symbol_19);
  return;
}

void import_level_inner(uint3 coord, const constant UBO* const tint_symbol_20, device Buffer* const tint_symbol_21, texture2d<float, access::sample> tint_symbol_22) {
  uint const offset = (coord[0] + (coord[1] * (*(tint_symbol_20)).width));
  (*(tint_symbol_21)).weights[offset] = tint_symbol_22.read(uint2(int2(uint3(coord).xy)), 0)[3];
}

kernel void import_level(const constant UBO* tint_symbol_23 [[buffer(2)]], device Buffer* tint_symbol_24 [[buffer(3)]], texture2d<float, access::sample> tint_symbol_25 [[texture(1)]], uint3 coord [[thread_position_in_grid]]) {
  import_level_inner(coord, tint_symbol_23, tint_symbol_24, tint_symbol_25);
  return;
}

void export_level_inner(uint3 coord, texture2d<float, access::write> tint_symbol_26, const constant UBO* const tint_symbol_27, const device Buffer* const tint_symbol_28, device Buffer* const tint_symbol_29) {
  if (all((uint3(coord).xy < uint2(int2(tint_symbol_26.get_width(), tint_symbol_26.get_height()))))) {
    uint const dst_offset = (coord[0] + (coord[1] * (*(tint_symbol_27)).width));
    uint const src_offset = ((coord[0] * 2u) + ((coord[1] * 2u) * (*(tint_symbol_27)).width));
    float const a_1 = (*(tint_symbol_28)).weights[(src_offset + 0u)];
    float const b = (*(tint_symbol_28)).weights[(src_offset + 1u)];
    float const c = (*(tint_symbol_28)).weights[((src_offset + 0u) + (*(tint_symbol_27)).width)];
    float const d = (*(tint_symbol_28)).weights[((src_offset + 1u) + (*(tint_symbol_27)).width)];
    float const sum = dot(float4(a_1, b, c, d), float4(1.0f));
    (*(tint_symbol_29)).weights[dst_offset] = (sum / 4.0f);
    float4 const probabilities = (float4(a_1, (a_1 + b), ((a_1 + b) + c), sum) / fmax(sum, 0.0001f));
    tint_symbol_26.write(probabilities, uint2(int2(uint3(coord).xy)));
  }
}

kernel void export_level(texture2d<float, access::write> tint_symbol_30 [[texture(2)]], const constant UBO* tint_symbol_31 [[buffer(2)]], const device Buffer* tint_symbol_32 [[buffer(4)]], device Buffer* tint_symbol_33 [[buffer(3)]], uint3 coord [[thread_position_in_grid]]) {
  export_level_inner(coord, tint_symbol_30, tint_symbol_31, tint_symbol_32, tint_symbol_33);
  return;
}