1#pragma clang diagnostic ignored "-Wmissing-prototypes" 2 3#include <metal_stdlib> 4#include <simd/simd.h> 5 6using namespace metal; 7 8struct SSBO 9{ 10 uint4 u; 11 int4 i; 12}; 13 14// Implementation of the GLSL findLSB() function 15template<typename T> 16inline T spvFindLSB(T x) 17{ 18 return select(ctz(x), T(-1), x == T(0)); 19} 20 21// Implementation of the signed GLSL findMSB() function 22template<typename T> 23inline T spvFindSMSB(T x) 24{ 25 T v = select(x, T(-1) - x, x < T(0)); 26 return select(clz(T(0)) - (clz(v) + T(1)), T(-1), v == T(0)); 27} 28 29// Implementation of the unsigned GLSL findMSB() function 30template<typename T> 31inline T spvFindUMSB(T x) 32{ 33 return select(clz(T(0)) - (clz(x) + T(1)), T(-1), x == T(0)); 34} 35 36kernel void main0(device SSBO& _4 [[buffer(0)]]) 37{ 38 uint4 _19 = _4.u; 39 int4 _20 = _4.i; 40 _4.u = spvFindLSB(_19); 41 _4.i = int4(spvFindLSB(_19)); 42 _4.u = uint4(spvFindLSB(_20)); 43 _4.i = spvFindLSB(_20); 44 _4.u = spvFindUMSB(_19); 45 _4.i = int4(spvFindUMSB(_19)); 46 _4.u = spvFindUMSB(uint4(_20)); 47 _4.i = int4(spvFindUMSB(uint4(_20))); 48 _4.u = uint4(spvFindSMSB(int4(_19))); 49 _4.i = spvFindSMSB(int4(_19)); 50 _4.u = uint4(spvFindSMSB(_20)); 51 _4.i = spvFindSMSB(_20); 52} 53 54