SPIRV-Cross/reference/shaders-msl/comp/barriers.comp
2019-09-24 14:35:25 -04:00

91 lines
2.0 KiB
Plaintext

#pragma clang diagnostic ignored "-Wmissing-prototypes"
#include <metal_stdlib>
#include <simd/simd.h>
using namespace metal;
constant uint3 gl_WorkGroupSize [[maybe_unused]] = uint3(4u, 1u, 1u);
static inline __attribute__((always_inline))
void barrier_shared()
{
threadgroup_barrier(mem_flags::mem_threadgroup);
}
static inline __attribute__((always_inline))
void full_barrier()
{
threadgroup_barrier(mem_flags::mem_device | mem_flags::mem_threadgroup | mem_flags::mem_texture);
}
static inline __attribute__((always_inline))
void image_barrier()
{
threadgroup_barrier(mem_flags::mem_texture);
}
static inline __attribute__((always_inline))
void buffer_barrier()
{
threadgroup_barrier(mem_flags::mem_device);
}
static inline __attribute__((always_inline))
void group_barrier()
{
threadgroup_barrier(mem_flags::mem_device | mem_flags::mem_threadgroup | mem_flags::mem_texture);
}
static inline __attribute__((always_inline))
void barrier_shared_exec()
{
threadgroup_barrier(mem_flags::mem_threadgroup);
}
static inline __attribute__((always_inline))
void full_barrier_exec()
{
threadgroup_barrier(mem_flags::mem_device | mem_flags::mem_threadgroup | mem_flags::mem_texture);
}
static inline __attribute__((always_inline))
void image_barrier_exec()
{
threadgroup_barrier(mem_flags::mem_texture);
}
static inline __attribute__((always_inline))
void buffer_barrier_exec()
{
threadgroup_barrier(mem_flags::mem_device);
}
static inline __attribute__((always_inline))
void group_barrier_exec()
{
threadgroup_barrier(mem_flags::mem_device | mem_flags::mem_threadgroup | mem_flags::mem_texture);
}
static inline __attribute__((always_inline))
void exec_barrier()
{
threadgroup_barrier(mem_flags::mem_threadgroup);
}
kernel void main0()
{
barrier_shared();
full_barrier();
image_barrier();
buffer_barrier();
group_barrier();
barrier_shared_exec();
full_barrier_exec();
image_barrier_exec();
buffer_barrier_exec();
group_barrier_exec();
exec_barrier();
}