diff options
author | Hans-Kristian Arntzen <post@arntzen-software.no> | 2022-01-18 14:39:16 +0300 |
---|---|---|
committer | Hans-Kristian Arntzen <post@arntzen-software.no> | 2022-01-18 14:39:16 +0300 |
commit | ac46140ba394ba5f85b6cfc72a5f5810084e8191 (patch) | |
tree | 2b2253b3486fd10b0126c9d4233a80114c1bdde9 /reference/shaders-msl-no-opt | |
parent | 48b5a9069f69b2370188a9a9ec4318cc784bb291 (diff) |
Test aliased names in declared LUTs.
Diffstat (limited to 'reference/shaders-msl-no-opt')
-rw-r--r-- | reference/shaders-msl-no-opt/asm/comp/constant-lut-name-aliasing.asm.comp | 61 |
1 files changed, 61 insertions, 0 deletions
diff --git a/reference/shaders-msl-no-opt/asm/comp/constant-lut-name-aliasing.asm.comp b/reference/shaders-msl-no-opt/asm/comp/constant-lut-name-aliasing.asm.comp new file mode 100644 index 00000000..a5b6fc32 --- /dev/null +++ b/reference/shaders-msl-no-opt/asm/comp/constant-lut-name-aliasing.asm.comp @@ -0,0 +1,61 @@ +#pragma clang diagnostic ignored "-Wmissing-prototypes" +#pragma clang diagnostic ignored "-Wmissing-braces" + +#include <metal_stdlib> +#include <simd/simd.h> + +using namespace metal; + +template<typename T, size_t Num> +struct spvUnsafeArray +{ + T elements[Num ? Num : 1]; + + thread T& operator [] (size_t pos) thread + { + return elements[pos]; + } + constexpr const thread T& operator [] (size_t pos) const thread + { + return elements[pos]; + } + + device T& operator [] (size_t pos) device + { + return elements[pos]; + } + constexpr const device T& operator [] (size_t pos) const device + { + return elements[pos]; + } + + constexpr const constant T& operator [] (size_t pos) const constant + { + return elements[pos]; + } + + threadgroup T& operator [] (size_t pos) threadgroup + { + return elements[pos]; + } + constexpr const threadgroup T& operator [] (size_t pos) const threadgroup + { + return elements[pos]; + } +}; + +struct SSBO +{ + int values[1]; +}; + +constant uint3 gl_WorkGroupSize [[maybe_unused]] = uint3(4u, 4u, 1u); + +constant spvUnsafeArray<int, 4> indexable = spvUnsafeArray<int, 4>({ 0, 1, 2, 3 }); +constant spvUnsafeArray<int, 4> indexable_1 = spvUnsafeArray<int, 4>({ 4, 5, 6, 7 }); + +kernel void main0(device SSBO& _6 [[buffer(0)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]], uint3 gl_LocalInvocationID [[thread_position_in_threadgroup]]) +{ + _6.values[gl_GlobalInvocationID.x] = indexable[gl_LocalInvocationID.x] + indexable_1[gl_LocalInvocationID.y]; +} + |