blob: f424448fdb97fd19f05bec67b4eac42fd60f31c6 [file] [log] [blame]
#include <metal_stdlib>
#include <simd/simd.h>
using namespace metal;
struct Alloc
{
uint offset;
};
struct Config
{
uint n_elements;
uint n_pathseg;
uint width_in_tiles;
uint height_in_tiles;
Alloc tile_alloc;
Alloc bin_alloc;
Alloc ptcl_alloc;
Alloc pathseg_alloc;
Alloc anno_alloc;
Alloc trans_alloc;
Alloc bbox_alloc;
uint n_trans;
uint trans_offset;
uint pathtag_offset;
uint linewidth_offset;
uint pathseg_offset;
};
struct ConfigBuf
{
Config conf;
};
struct Memory
{
uint mem_offset;
uint mem_error;
uint memory[1];
};
constant uint3 gl_WorkGroupSize [[maybe_unused]] = uint3(512u, 1u, 1u);
kernel void main0(device Memory& _44 [[buffer(0)]], const device ConfigBuf& _21 [[buffer(1)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
{
uint ix = gl_GlobalInvocationID.x;
if (ix < _21.conf.n_elements)
{
uint out_ix = (_21.conf.bbox_alloc.offset >> uint(2)) + (4u * ix);
_44.memory[out_ix] = 65535u;
_44.memory[out_ix + 1u] = 65535u;
_44.memory[out_ix + 2u] = 0u;
_44.memory[out_ix + 3u] = 0u;
}
}