mirror of https://github.com/bkaradzic/bgfx
Updated spirv-cross.
This commit is contained in:
Бранимир Караџић
parent
ab5cd6dc87
commit
0228691612
|
|
@ -1,7 +1,7 @@
|
|||
#!/bin/bash
|
||||
|
||||
GLSLANG_REV=4b5159ea8170fa34e29f13448fddebf88e0a722a
|
||||
SPIRV_TOOLS_REV=bbb29870b510f83f99994358179c9ea6838c3100
|
||||
GLSLANG_REV=5d0e01d9e7f9fae7609843d00f9fc5f41c3ddfdd
|
||||
SPIRV_TOOLS_REV=e0d5544c9864e1b6852c4ab949892898b2866c3e
|
||||
SPIRV_HEADERS_REV=601d738723ac381741311c6c98c36d6170be14a2
|
||||
|
||||
if [ -z $PROTOCOL ]; then
|
||||
|
|
|
|||
|
|
@ -1,69 +0,0 @@
|
|||
static const uint3 gl_WorkGroupSize = uint3(1u, 1u, 1u);
|
||||
|
||||
RWByteAddressBuffer _9 : register(u0, space0);
|
||||
|
||||
static uint4 gl_SubgroupEqMask;
|
||||
static uint4 gl_SubgroupGeMask;
|
||||
static uint4 gl_SubgroupGtMask;
|
||||
static uint4 gl_SubgroupLeMask;
|
||||
static uint4 gl_SubgroupLtMask;
|
||||
void comp_main()
|
||||
{
|
||||
_9.Store(0, asuint(float(WaveGetLaneCount())));
|
||||
_9.Store(0, asuint(float(WaveGetLaneIndex())));
|
||||
_9.Store(0, asuint(float4(gl_SubgroupEqMask).x));
|
||||
_9.Store(0, asuint(float4(gl_SubgroupGeMask).x));
|
||||
_9.Store(0, asuint(float4(gl_SubgroupGtMask).x));
|
||||
_9.Store(0, asuint(float4(gl_SubgroupLeMask).x));
|
||||
_9.Store(0, asuint(float4(gl_SubgroupLtMask).x));
|
||||
uint4 _75 = WaveActiveBallot(true);
|
||||
float4 _88 = WaveActiveSum(20.0f.xxxx);
|
||||
int4 _94 = WaveActiveSum(int4(20, 20, 20, 20));
|
||||
float4 _96 = WaveActiveProduct(20.0f.xxxx);
|
||||
int4 _98 = WaveActiveProduct(int4(20, 20, 20, 20));
|
||||
float4 _127 = WavePrefixProduct(_96) * _96;
|
||||
int4 _129 = WavePrefixProduct(_98) * _98;
|
||||
}
|
||||
|
||||
[numthreads(1, 1, 1)]
|
||||
void main()
|
||||
{
|
||||
gl_SubgroupEqMask = 1u << (WaveGetLaneIndex() - uint4(0, 32, 64, 96));
|
||||
if (WaveGetLaneIndex() >= 32) gl_SubgroupEqMask.x = 0;
|
||||
if (WaveGetLaneIndex() >= 64 || WaveGetLaneIndex() < 32) gl_SubgroupEqMask.y = 0;
|
||||
if (WaveGetLaneIndex() >= 96 || WaveGetLaneIndex() < 64) gl_SubgroupEqMask.z = 0;
|
||||
if (WaveGetLaneIndex() < 96) gl_SubgroupEqMask.w = 0;
|
||||
gl_SubgroupGeMask = ~((1u << (WaveGetLaneIndex() - uint4(0, 32, 64, 96))) - 1u);
|
||||
if (WaveGetLaneIndex() >= 32) gl_SubgroupGeMask.x = 0u;
|
||||
if (WaveGetLaneIndex() >= 64) gl_SubgroupGeMask.y = 0u;
|
||||
if (WaveGetLaneIndex() >= 96) gl_SubgroupGeMask.z = 0u;
|
||||
if (WaveGetLaneIndex() < 32) gl_SubgroupGeMask.y = ~0u;
|
||||
if (WaveGetLaneIndex() < 64) gl_SubgroupGeMask.z = ~0u;
|
||||
if (WaveGetLaneIndex() < 96) gl_SubgroupGeMask.w = ~0u;
|
||||
uint gt_lane_index = WaveGetLaneIndex() + 1;
|
||||
gl_SubgroupGtMask = ~((1u << (gt_lane_index - uint4(0, 32, 64, 96))) - 1u);
|
||||
if (gt_lane_index >= 32) gl_SubgroupGtMask.x = 0u;
|
||||
if (gt_lane_index >= 64) gl_SubgroupGtMask.y = 0u;
|
||||
if (gt_lane_index >= 96) gl_SubgroupGtMask.z = 0u;
|
||||
if (gt_lane_index >= 128) gl_SubgroupGtMask.w = 0u;
|
||||
if (gt_lane_index < 32) gl_SubgroupGtMask.y = ~0u;
|
||||
if (gt_lane_index < 64) gl_SubgroupGtMask.z = ~0u;
|
||||
if (gt_lane_index < 96) gl_SubgroupGtMask.w = ~0u;
|
||||
uint le_lane_index = WaveGetLaneIndex() + 1;
|
||||
gl_SubgroupLeMask = (1u << (le_lane_index - uint4(0, 32, 64, 96))) - 1u;
|
||||
if (le_lane_index >= 32) gl_SubgroupLeMask.x = ~0u;
|
||||
if (le_lane_index >= 64) gl_SubgroupLeMask.y = ~0u;
|
||||
if (le_lane_index >= 96) gl_SubgroupLeMask.z = ~0u;
|
||||
if (le_lane_index >= 128) gl_SubgroupLeMask.w = ~0u;
|
||||
if (le_lane_index < 32) gl_SubgroupLeMask.y = 0u;
|
||||
if (le_lane_index < 64) gl_SubgroupLeMask.z = 0u;
|
||||
if (le_lane_index < 96) gl_SubgroupLeMask.w = 0u;
|
||||
gl_SubgroupLtMask = (1u << (WaveGetLaneIndex() - uint4(0, 32, 64, 96))) - 1u;
|
||||
if (WaveGetLaneIndex() >= 32) gl_SubgroupLtMask.x = ~0u;
|
||||
if (WaveGetLaneIndex() >= 64) gl_SubgroupLtMask.y = ~0u;
|
||||
if (WaveGetLaneIndex() >= 96) gl_SubgroupLtMask.z = ~0u;
|
||||
if (WaveGetLaneIndex() < 32) gl_SubgroupLtMask.y = 0u;
|
||||
if (WaveGetLaneIndex() < 64) gl_SubgroupLtMask.z = 0u;
|
||||
if (WaveGetLaneIndex() < 96) gl_SubgroupLtMask.w = 0u;
|
||||
comp_main();
|
||||
}
|
||||
|
|
@ -1,45 +0,0 @@
|
|||
static min16float4 v4;
|
||||
static min16float3 v3;
|
||||
static min16float v1;
|
||||
static min16float2 v2;
|
||||
static float o1;
|
||||
static float2 o2;
|
||||
static float3 o3;
|
||||
static float4 o4;
|
||||
|
||||
struct SPIRV_Cross_Input
|
||||
{
|
||||
min16float v1 : TEXCOORD0;
|
||||
min16float2 v2 : TEXCOORD1;
|
||||
min16float3 v3 : TEXCOORD2;
|
||||
min16float4 v4 : TEXCOORD3;
|
||||
};
|
||||
|
||||
struct SPIRV_Cross_Output
|
||||
{
|
||||
float o1 : SV_Target0;
|
||||
float2 o2 : SV_Target1;
|
||||
float3 o3 : SV_Target2;
|
||||
float4 o4 : SV_Target3;
|
||||
};
|
||||
|
||||
void frag_main()
|
||||
{
|
||||
min16float4 _324;
|
||||
min16float4 _387 = modf(v4, _324);
|
||||
}
|
||||
|
||||
SPIRV_Cross_Output main(SPIRV_Cross_Input stage_input)
|
||||
{
|
||||
v4 = stage_input.v4;
|
||||
v3 = stage_input.v3;
|
||||
v1 = stage_input.v1;
|
||||
v2 = stage_input.v2;
|
||||
frag_main();
|
||||
SPIRV_Cross_Output stage_output;
|
||||
stage_output.o1 = o1;
|
||||
stage_output.o2 = o2;
|
||||
stage_output.o3 = o3;
|
||||
stage_output.o4 = o4;
|
||||
return stage_output;
|
||||
}
|
||||
|
|
@ -1,9 +0,0 @@
|
|||
#include <metal_stdlib>
|
||||
#include <simd/simd.h>
|
||||
|
||||
using namespace metal;
|
||||
|
||||
fragment void main0()
|
||||
{
|
||||
}
|
||||
|
||||
|
|
@ -1,99 +0,0 @@
|
|||
#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 VertexOutput
|
||||
{
|
||||
float4 pos;
|
||||
float2 uv;
|
||||
};
|
||||
|
||||
struct VertexOutput_1
|
||||
{
|
||||
float2 uv;
|
||||
};
|
||||
|
||||
struct HSOut
|
||||
{
|
||||
float2 uv;
|
||||
};
|
||||
|
||||
struct main0_out
|
||||
{
|
||||
HSOut _entryPointOutput;
|
||||
float4 gl_Position;
|
||||
};
|
||||
|
||||
struct main0_in
|
||||
{
|
||||
float2 VertexOutput_uv [[attribute(0)]];
|
||||
float4 gl_Position [[attribute(1)]];
|
||||
};
|
||||
|
||||
kernel void main0(main0_in in [[stage_in]], uint gl_InvocationID [[thread_index_in_threadgroup]], uint gl_PrimitiveID [[threadgroup_position_in_grid]], device main0_out* spvOut [[buffer(28)]], constant uint* spvIndirectParams [[buffer(29)]], device MTLTriangleTessellationFactorsHalf* spvTessLevel [[buffer(26)]], threadgroup main0_in* gl_in [[threadgroup(0)]])
|
||||
{
|
||||
device main0_out* gl_out = &spvOut[gl_PrimitiveID * 3];
|
||||
if (gl_InvocationID < spvIndirectParams[0])
|
||||
gl_in[gl_InvocationID] = in;
|
||||
threadgroup_barrier(mem_flags::mem_threadgroup);
|
||||
if (gl_InvocationID >= 3)
|
||||
return;
|
||||
spvUnsafeArray<VertexOutput, 3> _223 = spvUnsafeArray<VertexOutput, 3>({ VertexOutput{ gl_in[0].gl_Position, gl_in[0].VertexOutput_uv }, VertexOutput{ gl_in[1].gl_Position, gl_in[1].VertexOutput_uv }, VertexOutput{ gl_in[2].gl_Position, gl_in[2].VertexOutput_uv } });
|
||||
spvUnsafeArray<VertexOutput, 3> param;
|
||||
param = _223;
|
||||
gl_out[gl_InvocationID].gl_Position = param[gl_InvocationID].pos;
|
||||
gl_out[gl_InvocationID]._entryPointOutput.uv = param[gl_InvocationID].uv;
|
||||
threadgroup_barrier(mem_flags::mem_device | mem_flags::mem_threadgroup);
|
||||
if (int(gl_InvocationID) == 0)
|
||||
{
|
||||
float2 _174 = float2(1.0) + gl_in[0].VertexOutput_uv;
|
||||
float _175 = _174.x;
|
||||
spvTessLevel[gl_PrimitiveID].edgeTessellationFactor[0] = half(_175);
|
||||
spvTessLevel[gl_PrimitiveID].edgeTessellationFactor[1] = half(_175);
|
||||
spvTessLevel[gl_PrimitiveID].edgeTessellationFactor[2] = half(_175);
|
||||
spvTessLevel[gl_PrimitiveID].insideTessellationFactor = half(_175);
|
||||
}
|
||||
}
|
||||
|
||||
|
|
@ -1,89 +0,0 @@
|
|||
#include <metal_stdlib>
|
||||
#include <simd/simd.h>
|
||||
|
||||
using namespace metal;
|
||||
|
||||
struct Baz
|
||||
{
|
||||
int e;
|
||||
int f;
|
||||
};
|
||||
|
||||
struct Foo
|
||||
{
|
||||
int a;
|
||||
int b;
|
||||
};
|
||||
|
||||
struct Bar
|
||||
{
|
||||
int c;
|
||||
int d;
|
||||
};
|
||||
|
||||
constant uint3 gl_WorkGroupSize [[maybe_unused]] = uint3(3u, 3u, 2u);
|
||||
|
||||
struct spvDescriptorSetBuffer0
|
||||
{
|
||||
constant Foo* m_34 [[id(0)]];
|
||||
constant Bar* m_40 [[id(1)]];
|
||||
};
|
||||
|
||||
struct spvDescriptorSetBuffer1
|
||||
{
|
||||
device Baz* baz [[id(0)]][3][3][2];
|
||||
};
|
||||
|
||||
kernel void main0(constant spvDescriptorSetBuffer0& spvDescriptorSet0 [[buffer(0)]], constant spvDescriptorSetBuffer1& spvDescriptorSet1 [[buffer(1)]], constant uint* spvDynamicOffsets [[buffer(23)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
|
||||
{
|
||||
constant auto& _34 = *(constant Foo* )((constant char* )spvDescriptorSet0.m_34 + spvDynamicOffsets[0]);
|
||||
device Baz* baz[3][3][2] =
|
||||
{
|
||||
{
|
||||
{
|
||||
(device Baz* )((device char* )spvDescriptorSet1.baz[0][0][0] + spvDynamicOffsets[1]),
|
||||
(device Baz* )((device char* )spvDescriptorSet1.baz[0][0][1] + spvDynamicOffsets[2]),
|
||||
},
|
||||
{
|
||||
(device Baz* )((device char* )spvDescriptorSet1.baz[0][1][0] + spvDynamicOffsets[3]),
|
||||
(device Baz* )((device char* )spvDescriptorSet1.baz[0][1][1] + spvDynamicOffsets[4]),
|
||||
},
|
||||
{
|
||||
(device Baz* )((device char* )spvDescriptorSet1.baz[0][2][0] + spvDynamicOffsets[5]),
|
||||
(device Baz* )((device char* )spvDescriptorSet1.baz[0][2][1] + spvDynamicOffsets[6]),
|
||||
},
|
||||
},
|
||||
{
|
||||
{
|
||||
(device Baz* )((device char* )spvDescriptorSet1.baz[1][0][0] + spvDynamicOffsets[7]),
|
||||
(device Baz* )((device char* )spvDescriptorSet1.baz[1][0][1] + spvDynamicOffsets[8]),
|
||||
},
|
||||
{
|
||||
(device Baz* )((device char* )spvDescriptorSet1.baz[1][1][0] + spvDynamicOffsets[9]),
|
||||
(device Baz* )((device char* )spvDescriptorSet1.baz[1][1][1] + spvDynamicOffsets[10]),
|
||||
},
|
||||
{
|
||||
(device Baz* )((device char* )spvDescriptorSet1.baz[1][2][0] + spvDynamicOffsets[11]),
|
||||
(device Baz* )((device char* )spvDescriptorSet1.baz[1][2][1] + spvDynamicOffsets[12]),
|
||||
},
|
||||
},
|
||||
{
|
||||
{
|
||||
(device Baz* )((device char* )spvDescriptorSet1.baz[2][0][0] + spvDynamicOffsets[13]),
|
||||
(device Baz* )((device char* )spvDescriptorSet1.baz[2][0][1] + spvDynamicOffsets[14]),
|
||||
},
|
||||
{
|
||||
(device Baz* )((device char* )spvDescriptorSet1.baz[2][1][0] + spvDynamicOffsets[15]),
|
||||
(device Baz* )((device char* )spvDescriptorSet1.baz[2][1][1] + spvDynamicOffsets[16]),
|
||||
},
|
||||
{
|
||||
(device Baz* )((device char* )spvDescriptorSet1.baz[2][2][0] + spvDynamicOffsets[17]),
|
||||
(device Baz* )((device char* )spvDescriptorSet1.baz[2][2][1] + spvDynamicOffsets[18]),
|
||||
},
|
||||
},
|
||||
};
|
||||
|
||||
baz[gl_GlobalInvocationID.x][gl_GlobalInvocationID.y][gl_GlobalInvocationID.z]->e = _34.a + (*spvDescriptorSet0.m_40).c;
|
||||
baz[gl_GlobalInvocationID.x][gl_GlobalInvocationID.y][gl_GlobalInvocationID.z]->f = _34.b * (*spvDescriptorSet0.m_40).d;
|
||||
}
|
||||
|
||||
|
|
@ -1,24 +0,0 @@
|
|||
#include <metal_stdlib>
|
||||
#include <simd/simd.h>
|
||||
|
||||
using namespace metal;
|
||||
|
||||
struct SSBO0
|
||||
{
|
||||
short4 inputs[1];
|
||||
};
|
||||
|
||||
struct SSBO1
|
||||
{
|
||||
int4 outputs[1];
|
||||
};
|
||||
|
||||
constant uint3 gl_WorkGroupSize [[maybe_unused]] = uint3(1u);
|
||||
|
||||
kernel void main0(device SSBO0& _25 [[buffer(0)]], device SSBO1& _39 [[buffer(1)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
|
||||
{
|
||||
_39.outputs[gl_GlobalInvocationID.x].x = int(as_type<uint>(as_type<half2>(_25.inputs[gl_GlobalInvocationID.x].xy) + half2(half(1.0))));
|
||||
_39.outputs[gl_GlobalInvocationID.x].y = as_type<int>(_25.inputs[gl_GlobalInvocationID.x].zw);
|
||||
_39.outputs[gl_GlobalInvocationID.x].z = int(as_type<uint>(ushort2(_25.inputs[gl_GlobalInvocationID.x].xy)));
|
||||
}
|
||||
|
||||
|
|
@ -1,30 +0,0 @@
|
|||
#include <metal_stdlib>
|
||||
#include <simd/simd.h>
|
||||
|
||||
using namespace metal;
|
||||
|
||||
struct SSBO1
|
||||
{
|
||||
short4 outputs[1];
|
||||
};
|
||||
|
||||
struct SSBO0
|
||||
{
|
||||
int4 inputs[1];
|
||||
};
|
||||
|
||||
struct UBO
|
||||
{
|
||||
half4 const0;
|
||||
};
|
||||
|
||||
constant uint3 gl_WorkGroupSize [[maybe_unused]] = uint3(1u);
|
||||
|
||||
kernel void main0(device SSBO1& _21 [[buffer(0)]], device SSBO0& _29 [[buffer(1)]], constant UBO& _40 [[buffer(2)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
|
||||
{
|
||||
short2 _47 = as_type<short2>(_29.inputs[gl_GlobalInvocationID.x].x) + as_type<short2>(_40.const0.xy);
|
||||
_21.outputs[gl_GlobalInvocationID.x] = short4(_47.x, _47.y, _21.outputs[gl_GlobalInvocationID.x].z, _21.outputs[gl_GlobalInvocationID.x].w);
|
||||
short2 _66 = short2(as_type<ushort2>(uint(_29.inputs[gl_GlobalInvocationID.x].y)) - as_type<ushort2>(_40.const0.zw));
|
||||
_21.outputs[gl_GlobalInvocationID.x] = short4(_21.outputs[gl_GlobalInvocationID.x].x, _21.outputs[gl_GlobalInvocationID.x].y, _66.x, _66.y);
|
||||
}
|
||||
|
||||
|
|
@ -1,26 +0,0 @@
|
|||
#include <metal_stdlib>
|
||||
#include <simd/simd.h>
|
||||
|
||||
using namespace metal;
|
||||
|
||||
struct SSBO
|
||||
{
|
||||
int s32;
|
||||
uint u32;
|
||||
};
|
||||
|
||||
constant uint3 gl_WorkGroupSize [[maybe_unused]] = uint3(1u);
|
||||
|
||||
constant long _164 = {};
|
||||
|
||||
kernel void main0(device SSBO& _96 [[buffer(0)]])
|
||||
{
|
||||
long4 _139;
|
||||
ulong4 _143;
|
||||
_139 = abs((_139 + long4(30l, 40l, 50l, 60l)) + long4(_143 + ulong4(999999999999999999ul, 8888888888888888ul, 77777777777777777ul, 6666666666666666ul)));
|
||||
_143 += ulong4(long4(999999999999999999l, 8888888888888888l, 77777777777777777l, 6666666666666666l));
|
||||
ulong _109 = ulong(_164);
|
||||
_96.s32 = int(uint(((ulong(_139.x) + _143.y) + _109) + _109));
|
||||
_96.u32 = uint(((ulong(_139.y) + _143.z) + ulong(_164 + 1l)) + _109);
|
||||
}
|
||||
|
||||
|
|
@ -1,16 +0,0 @@
|
|||
#include <metal_stdlib>
|
||||
#include <simd/simd.h>
|
||||
|
||||
using namespace metal;
|
||||
|
||||
struct main0_in
|
||||
{
|
||||
half4 v4 [[user(locn3)]];
|
||||
};
|
||||
|
||||
fragment void main0(main0_in in [[stage_in]])
|
||||
{
|
||||
half4 _491;
|
||||
half4 _563 = modf(in.v4, _491);
|
||||
}
|
||||
|
||||
|
|
@ -1,27 +0,0 @@
|
|||
#include <metal_stdlib>
|
||||
#include <simd/simd.h>
|
||||
|
||||
using namespace metal;
|
||||
|
||||
struct main0_out
|
||||
{
|
||||
float FragColor [[color(0)]];
|
||||
};
|
||||
|
||||
struct main0_in
|
||||
{
|
||||
float3 vUV [[user(locn0)]];
|
||||
};
|
||||
|
||||
fragment main0_out main0(main0_in in [[stage_in]], depth2d<float> uTex [[texture(0)]], depth2d<float> uSampler [[texture(1)]], sampler uSamp [[sampler(0)]], sampler uSamplerSmplr [[sampler(1)]])
|
||||
{
|
||||
main0_out out = {};
|
||||
out.FragColor = uSampler.sample_compare(uSamplerSmplr, in.vUV.xy, in.vUV.z);
|
||||
out.FragColor += uTex.sample_compare(uSamp, in.vUV.xy, in.vUV.z);
|
||||
out.FragColor += uTex.sample_compare(uSamp, in.vUV.xy, in.vUV.z);
|
||||
out.FragColor += uSampler.sample_compare(uSamplerSmplr, in.vUV.xy, in.vUV.z);
|
||||
out.FragColor += uTex.sample_compare(uSamp, in.vUV.xy, in.vUV.z);
|
||||
out.FragColor += uSampler.sample_compare(uSamplerSmplr, in.vUV.xy, in.vUV.z);
|
||||
return out;
|
||||
}
|
||||
|
||||
|
|
@ -1,94 +0,0 @@
|
|||
#pragma clang diagnostic ignored "-Wmissing-prototypes"
|
||||
|
||||
#include <metal_stdlib>
|
||||
#include <simd/simd.h>
|
||||
|
||||
using namespace metal;
|
||||
|
||||
struct SSBO
|
||||
{
|
||||
float FragColor;
|
||||
};
|
||||
|
||||
constant uint3 gl_WorkGroupSize [[maybe_unused]] = uint3(1u);
|
||||
|
||||
inline uint4 spvSubgroupBallot(bool value)
|
||||
{
|
||||
simd_vote vote = simd_ballot(value);
|
||||
// simd_ballot() returns a 64-bit integer-like object, but
|
||||
// SPIR-V callers expect a uint4. We must convert.
|
||||
// FIXME: This won't include higher bits if Apple ever supports
|
||||
// 128 lanes in an SIMD-group.
|
||||
return uint4((uint)((simd_vote::vote_t)vote & 0xFFFFFFFF), (uint)(((simd_vote::vote_t)vote >> 32) & 0xFFFFFFFF), 0, 0);
|
||||
}
|
||||
|
||||
inline bool spvSubgroupBallotBitExtract(uint4 ballot, uint bit)
|
||||
{
|
||||
return !!extract_bits(ballot[bit / 32], bit % 32, 1);
|
||||
}
|
||||
|
||||
inline uint spvSubgroupBallotFindLSB(uint4 ballot)
|
||||
{
|
||||
return select(ctz(ballot.x), select(32 + ctz(ballot.y), select(64 + ctz(ballot.z), select(96 + ctz(ballot.w), uint(-1), ballot.w == 0), ballot.z == 0), ballot.y == 0), ballot.x == 0);
|
||||
}
|
||||
|
||||
inline uint spvSubgroupBallotFindMSB(uint4 ballot)
|
||||
{
|
||||
return select(128 - (clz(ballot.w) + 1), select(96 - (clz(ballot.z) + 1), select(64 - (clz(ballot.y) + 1), select(32 - (clz(ballot.x) + 1), uint(-1), ballot.x == 0), ballot.y == 0), ballot.z == 0), ballot.w == 0);
|
||||
}
|
||||
|
||||
inline uint spvSubgroupBallotBitCount(uint4 ballot)
|
||||
{
|
||||
return popcount(ballot.x) + popcount(ballot.y) + popcount(ballot.z) + popcount(ballot.w);
|
||||
}
|
||||
|
||||
inline uint spvSubgroupBallotInclusiveBitCount(uint4 ballot, uint gl_SubgroupInvocationID)
|
||||
{
|
||||
uint4 mask = uint4(extract_bits(0xFFFFFFFF, 0, min(gl_SubgroupInvocationID + 1, 32u)), extract_bits(0xFFFFFFFF, 0, (uint)max((int)gl_SubgroupInvocationID + 1 - 32, 0)), uint2(0));
|
||||
return spvSubgroupBallotBitCount(ballot & mask);
|
||||
}
|
||||
|
||||
inline uint spvSubgroupBallotExclusiveBitCount(uint4 ballot, uint gl_SubgroupInvocationID)
|
||||
{
|
||||
uint4 mask = uint4(extract_bits(0xFFFFFFFF, 0, min(gl_SubgroupInvocationID, 32u)), extract_bits(0xFFFFFFFF, 0, (uint)max((int)gl_SubgroupInvocationID - 32, 0)), uint2(0));
|
||||
return spvSubgroupBallotBitCount(ballot & mask);
|
||||
}
|
||||
|
||||
template<typename T>
|
||||
inline bool spvSubgroupAllEqual(T value)
|
||||
{
|
||||
return simd_all(value == simd_broadcast_first(value));
|
||||
}
|
||||
|
||||
template<>
|
||||
inline bool spvSubgroupAllEqual(bool value)
|
||||
{
|
||||
return simd_all(value) || !simd_any(value);
|
||||
}
|
||||
|
||||
kernel void main0(device SSBO& _9 [[buffer(0)]], uint gl_NumSubgroups [[simdgroups_per_threadgroup]], uint gl_SubgroupID [[simdgroup_index_in_threadgroup]], uint gl_SubgroupSize [[thread_execution_width]], uint gl_SubgroupInvocationID [[thread_index_in_simdgroup]])
|
||||
{
|
||||
uint4 gl_SubgroupEqMask = gl_SubgroupInvocationID > 32 ? uint4(0, (1 << (gl_SubgroupInvocationID - 32)), uint2(0)) : uint4(1 << gl_SubgroupInvocationID, uint3(0));
|
||||
uint4 gl_SubgroupGeMask = uint4(extract_bits(0xFFFFFFFF, min(gl_SubgroupInvocationID, 32u), (uint)max(min((int)gl_SubgroupSize, 32) - (int)gl_SubgroupInvocationID, 0)), extract_bits(0xFFFFFFFF, (uint)max((int)gl_SubgroupInvocationID - 32, 0), (uint)max((int)gl_SubgroupSize - (int)max(gl_SubgroupInvocationID, 32u), 0)), uint2(0));
|
||||
uint4 gl_SubgroupGtMask = uint4(extract_bits(0xFFFFFFFF, min(gl_SubgroupInvocationID + 1, 32u), (uint)max(min((int)gl_SubgroupSize, 32) - (int)gl_SubgroupInvocationID - 1, 0)), extract_bits(0xFFFFFFFF, (uint)max((int)gl_SubgroupInvocationID + 1 - 32, 0), (uint)max((int)gl_SubgroupSize - (int)max(gl_SubgroupInvocationID + 1, 32u), 0)), uint2(0));
|
||||
uint4 gl_SubgroupLeMask = uint4(extract_bits(0xFFFFFFFF, 0, min(gl_SubgroupInvocationID + 1, 32u)), extract_bits(0xFFFFFFFF, 0, (uint)max((int)gl_SubgroupInvocationID + 1 - 32, 0)), uint2(0));
|
||||
uint4 gl_SubgroupLtMask = uint4(extract_bits(0xFFFFFFFF, 0, min(gl_SubgroupInvocationID, 32u)), extract_bits(0xFFFFFFFF, 0, (uint)max((int)gl_SubgroupInvocationID - 32, 0)), uint2(0));
|
||||
_9.FragColor = float(gl_NumSubgroups);
|
||||
_9.FragColor = float(gl_SubgroupID);
|
||||
_9.FragColor = float(gl_SubgroupSize);
|
||||
_9.FragColor = float(gl_SubgroupInvocationID);
|
||||
simdgroup_barrier(mem_flags::mem_device | mem_flags::mem_threadgroup | mem_flags::mem_texture);
|
||||
simdgroup_barrier(mem_flags::mem_device | mem_flags::mem_threadgroup | mem_flags::mem_texture);
|
||||
simdgroup_barrier(mem_flags::mem_device);
|
||||
simdgroup_barrier(mem_flags::mem_threadgroup);
|
||||
simdgroup_barrier(mem_flags::mem_texture);
|
||||
_9.FragColor = float4(gl_SubgroupEqMask).x;
|
||||
_9.FragColor = float4(gl_SubgroupGeMask).x;
|
||||
_9.FragColor = float4(gl_SubgroupGtMask).x;
|
||||
_9.FragColor = float4(gl_SubgroupLeMask).x;
|
||||
_9.FragColor = float4(gl_SubgroupLtMask).x;
|
||||
uint4 _83 = spvSubgroupBallot(true);
|
||||
float4 _165 = simd_prefix_inclusive_product(simd_product(float4(20.0)));
|
||||
int4 _167 = simd_prefix_inclusive_product(simd_product(int4(20)));
|
||||
}
|
||||
|
||||
|
|
@ -1,25 +0,0 @@
|
|||
#include <metal_stdlib>
|
||||
#include <simd/simd.h>
|
||||
|
||||
using namespace metal;
|
||||
|
||||
struct SSBO
|
||||
{
|
||||
float FragColor;
|
||||
};
|
||||
|
||||
constant uint3 gl_WorkGroupSize [[maybe_unused]] = uint3(1u);
|
||||
|
||||
kernel void main0(device SSBO& _9 [[buffer(0)]], uint gl_NumSubgroups [[quadgroups_per_threadgroup]], uint gl_SubgroupID [[quadgroup_index_in_threadgroup]], uint gl_SubgroupSize [[thread_execution_width]], uint gl_SubgroupInvocationID [[thread_index_in_quadgroup]])
|
||||
{
|
||||
_9.FragColor = float(gl_NumSubgroups);
|
||||
_9.FragColor = float(gl_SubgroupID);
|
||||
_9.FragColor = float(gl_SubgroupSize);
|
||||
_9.FragColor = float(gl_SubgroupInvocationID);
|
||||
simdgroup_barrier(mem_flags::mem_device | mem_flags::mem_threadgroup | mem_flags::mem_texture);
|
||||
simdgroup_barrier(mem_flags::mem_device | mem_flags::mem_threadgroup | mem_flags::mem_texture);
|
||||
simdgroup_barrier(mem_flags::mem_device);
|
||||
simdgroup_barrier(mem_flags::mem_threadgroup);
|
||||
simdgroup_barrier(mem_flags::mem_texture);
|
||||
}
|
||||
|
||||
|
|
@ -1,89 +0,0 @@
|
|||
#pragma clang diagnostic ignored "-Wmissing-prototypes"
|
||||
|
||||
#include <metal_stdlib>
|
||||
#include <simd/simd.h>
|
||||
|
||||
using namespace metal;
|
||||
|
||||
struct main0_out
|
||||
{
|
||||
float FragColor [[color(0)]];
|
||||
};
|
||||
|
||||
inline uint4 spvSubgroupBallot(bool value)
|
||||
{
|
||||
simd_vote vote = simd_ballot(value);
|
||||
// simd_ballot() returns a 64-bit integer-like object, but
|
||||
// SPIR-V callers expect a uint4. We must convert.
|
||||
// FIXME: This won't include higher bits if Apple ever supports
|
||||
// 128 lanes in an SIMD-group.
|
||||
return uint4((uint)((simd_vote::vote_t)vote & 0xFFFFFFFF), (uint)(((simd_vote::vote_t)vote >> 32) & 0xFFFFFFFF), 0, 0);
|
||||
}
|
||||
|
||||
inline bool spvSubgroupBallotBitExtract(uint4 ballot, uint bit)
|
||||
{
|
||||
return !!extract_bits(ballot[bit / 32], bit % 32, 1);
|
||||
}
|
||||
|
||||
inline uint spvSubgroupBallotFindLSB(uint4 ballot)
|
||||
{
|
||||
return select(ctz(ballot.x), select(32 + ctz(ballot.y), select(64 + ctz(ballot.z), select(96 + ctz(ballot.w), uint(-1), ballot.w == 0), ballot.z == 0), ballot.y == 0), ballot.x == 0);
|
||||
}
|
||||
|
||||
inline uint spvSubgroupBallotFindMSB(uint4 ballot)
|
||||
{
|
||||
return select(128 - (clz(ballot.w) + 1), select(96 - (clz(ballot.z) + 1), select(64 - (clz(ballot.y) + 1), select(32 - (clz(ballot.x) + 1), uint(-1), ballot.x == 0), ballot.y == 0), ballot.z == 0), ballot.w == 0);
|
||||
}
|
||||
|
||||
inline uint spvSubgroupBallotBitCount(uint4 ballot)
|
||||
{
|
||||
return popcount(ballot.x) + popcount(ballot.y) + popcount(ballot.z) + popcount(ballot.w);
|
||||
}
|
||||
|
||||
inline uint spvSubgroupBallotInclusiveBitCount(uint4 ballot, uint gl_SubgroupInvocationID)
|
||||
{
|
||||
uint4 mask = uint4(extract_bits(0xFFFFFFFF, 0, min(gl_SubgroupInvocationID + 1, 32u)), extract_bits(0xFFFFFFFF, 0, (uint)max((int)gl_SubgroupInvocationID + 1 - 32, 0)), uint2(0));
|
||||
return spvSubgroupBallotBitCount(ballot & mask);
|
||||
}
|
||||
|
||||
inline uint spvSubgroupBallotExclusiveBitCount(uint4 ballot, uint gl_SubgroupInvocationID)
|
||||
{
|
||||
uint4 mask = uint4(extract_bits(0xFFFFFFFF, 0, min(gl_SubgroupInvocationID, 32u)), extract_bits(0xFFFFFFFF, 0, (uint)max((int)gl_SubgroupInvocationID - 32, 0)), uint2(0));
|
||||
return spvSubgroupBallotBitCount(ballot & mask);
|
||||
}
|
||||
|
||||
template<typename T>
|
||||
inline bool spvSubgroupAllEqual(T value)
|
||||
{
|
||||
return simd_all(value == simd_broadcast_first(value));
|
||||
}
|
||||
|
||||
template<>
|
||||
inline bool spvSubgroupAllEqual(bool value)
|
||||
{
|
||||
return simd_all(value) || !simd_any(value);
|
||||
}
|
||||
|
||||
fragment main0_out main0()
|
||||
{
|
||||
main0_out out = {};
|
||||
uint gl_SubgroupSize = simd_sum(1);
|
||||
uint gl_SubgroupInvocationID = simd_prefix_exclusive_sum(1);
|
||||
uint4 gl_SubgroupEqMask = gl_SubgroupInvocationID > 32 ? uint4(0, (1 << (gl_SubgroupInvocationID - 32)), uint2(0)) : uint4(1 << gl_SubgroupInvocationID, uint3(0));
|
||||
uint4 gl_SubgroupGeMask = uint4(extract_bits(0xFFFFFFFF, min(gl_SubgroupInvocationID, 32u), (uint)max(min((int)gl_SubgroupSize, 32) - (int)gl_SubgroupInvocationID, 0)), extract_bits(0xFFFFFFFF, (uint)max((int)gl_SubgroupInvocationID - 32, 0), (uint)max((int)gl_SubgroupSize - (int)max(gl_SubgroupInvocationID, 32u), 0)), uint2(0));
|
||||
uint4 gl_SubgroupGtMask = uint4(extract_bits(0xFFFFFFFF, min(gl_SubgroupInvocationID + 1, 32u), (uint)max(min((int)gl_SubgroupSize, 32) - (int)gl_SubgroupInvocationID - 1, 0)), extract_bits(0xFFFFFFFF, (uint)max((int)gl_SubgroupInvocationID + 1 - 32, 0), (uint)max((int)gl_SubgroupSize - (int)max(gl_SubgroupInvocationID + 1, 32u), 0)), uint2(0));
|
||||
uint4 gl_SubgroupLeMask = uint4(extract_bits(0xFFFFFFFF, 0, min(gl_SubgroupInvocationID + 1, 32u)), extract_bits(0xFFFFFFFF, 0, (uint)max((int)gl_SubgroupInvocationID + 1 - 32, 0)), uint2(0));
|
||||
uint4 gl_SubgroupLtMask = uint4(extract_bits(0xFFFFFFFF, 0, min(gl_SubgroupInvocationID, 32u)), extract_bits(0xFFFFFFFF, 0, (uint)max((int)gl_SubgroupInvocationID - 32, 0)), uint2(0));
|
||||
out.FragColor = float(gl_SubgroupSize);
|
||||
out.FragColor = float(gl_SubgroupInvocationID);
|
||||
out.FragColor = float4(gl_SubgroupEqMask).x;
|
||||
out.FragColor = float4(gl_SubgroupGeMask).x;
|
||||
out.FragColor = float4(gl_SubgroupGtMask).x;
|
||||
out.FragColor = float4(gl_SubgroupLeMask).x;
|
||||
out.FragColor = float4(gl_SubgroupLtMask).x;
|
||||
uint4 _63 = spvSubgroupBallot(true);
|
||||
float4 _147 = simd_prefix_inclusive_product(simd_product(float4(20.0)));
|
||||
int4 _149 = simd_prefix_inclusive_product(simd_product(int4(20)));
|
||||
return out;
|
||||
}
|
||||
|
||||
|
|
@ -261,12 +261,6 @@ struct main0_in
|
|||
float4 in_var_TEXCOORD0_0 [[user(locn4)]];
|
||||
};
|
||||
|
||||
static inline __attribute__((always_inline))
|
||||
void _353()
|
||||
{
|
||||
discard_fragment();
|
||||
}
|
||||
|
||||
fragment main0_out main0(main0_in in [[stage_in]], constant type_View& View [[buffer(0)]], constant type_PrimitiveDither& PrimitiveDither [[buffer(1)]], constant type_PrimitiveFade& PrimitiveFade [[buffer(2)]], constant type_Material& Material [[buffer(3)]], texture2d<float> Material_Texture2D_0 [[texture(0)]], texture2d<float> Material_Texture2D_3 [[texture(1)]], sampler Material_Texture2D_0Sampler [[sampler(0)]], sampler Material_Texture2D_3Sampler [[sampler(1)]], float4 gl_FragCoord [[position]])
|
||||
{
|
||||
main0_out out = {};
|
||||
|
|
@ -340,13 +334,13 @@ fragment main0_out main0(main0_in in [[stage_in]], constant type_View& View [[bu
|
|||
float _317 = fract(cos(dot(floor(gl_FragCoord.xy), float2(347.834503173828125, 3343.28369140625))) * 1000.0);
|
||||
if ((float((PrimitiveDither.PrimitiveDither_LODFactor < 0.0) ? ((PrimitiveDither.PrimitiveDither_LODFactor + 1.0) > _317) : (PrimitiveDither.PrimitiveDither_LODFactor < _317)) - 0.001000000047497451305389404296875) < 0.0)
|
||||
{
|
||||
_353();
|
||||
discard_fragment();
|
||||
}
|
||||
}
|
||||
}
|
||||
if ((((_218.z + ((fast::min(fast::max(1.0 - (_218.x * Material.Material_ScalarExpressions[2].y), 0.0), 1.0) + ((_237 + (_242.x * Material.Material_ScalarExpressions[2].z)) * 0.16666667163372039794921875)) + (-0.5))) * ((fast::clamp((View.View_RealTime * PrimitiveFade.PrimitiveFade_FadeTimeScaleBias.x) + PrimitiveFade.PrimitiveFade_FadeTimeScaleBias.y, 0.0, 1.0) + ((_237 + _254.x) * 0.16666667163372039794921875)) + (-0.5))) - 0.33329999446868896484375) < 0.0)
|
||||
{
|
||||
_353();
|
||||
discard_fragment();
|
||||
}
|
||||
float2 _351 = ((((in.in_var_TEXCOORD6.xy / float2(in.in_var_TEXCOORD6.w)) - View.View_TemporalAAJitter.xy) - ((in.in_var_TEXCOORD7.xy / float2(in.in_var_TEXCOORD7.w)) - View.View_TemporalAAJitter.zw)) * float2(0.2495000064373016357421875)) + float2(0.49999237060546875);
|
||||
out.gl_FragDepth = fast::min(_140.z / (_140.w + (sqrt(dot(_272, _272)) / (fast::max(sqrt(dot(_276, _276)) / sqrt(dot(_279, _279)), sqrt(dot(_284, _284)) / sqrt(dot(_287, _287))) / abs(dot(float3x3(View.View_ViewToTranslatedWorld[0].xyz, View.View_ViewToTranslatedWorld[1].xyz, View.View_ViewToTranslatedWorld[2].xyz) * float3(0.0, 0.0, 1.0), _151))))), gl_FragCoord.z);
|
||||
|
|
|
|||
|
|
@ -180,7 +180,7 @@ fragment main0_out main0(main0_in in [[stage_in]], constant type_Globals& _Globa
|
|||
float _665 = _633 + (((-_660.y) * _Globals.WhiteTint) * 0.0500000007450580596923828125);
|
||||
float _669 = _644 + ((_660.x * _Globals.WhiteTint) * 0.0500000007450580596923828125);
|
||||
float _674 = ((2.0 * _665) - (8.0 * _669)) + 4.0;
|
||||
float2 _680 = select(float2(_616, ((((-3.0) * _616) * _616) + (2.86999988555908203125 * _616)) - 0.2750000059604644775390625), _653, bool2(_Globals.WhiteTemp < 4000.0)) + (float2((3.0 * _665) / _674, (2.0 * _669) / _674) - _653);
|
||||
float2 _680 = select(float2(_616, (_616 * (((-3.0) * _616) + 2.86999988555908203125)) - 0.2750000059604644775390625), _653, bool2(_Globals.WhiteTemp < 4000.0)) + (float2((3.0 * _665) / _674, (2.0 * _669) / _674) - _653);
|
||||
float _681 = _680.x;
|
||||
float _682 = _680.y;
|
||||
float _683 = fast::max(_682, 1.0000000133514319600180897396058e-10);
|
||||
|
|
|
|||
|
|
@ -102,7 +102,7 @@ constant spvUnsafeArray<float, 10> _506 = spvUnsafeArray<float, 10>({ -2.3010299
|
|||
constant spvUnsafeArray<float, 10> _507 = spvUnsafeArray<float, 10>({ 0.801995217800140380859375, 1.19800484180450439453125, 1.5943000316619873046875, 1.99730002880096435546875, 2.3782999515533447265625, 2.7683999538421630859375, 3.0515000820159912109375, 3.2746293544769287109375, 3.32743072509765625, 3.32743072509765625 });
|
||||
|
||||
constant float3 _523 = {};
|
||||
constant float3 _3123 = {};
|
||||
constant float3 _3124 = {};
|
||||
|
||||
struct main0_out
|
||||
{
|
||||
|
|
@ -182,7 +182,7 @@ fragment main0_out main0(main0_in in [[stage_in]], constant type_Globals& _Globa
|
|||
float _691 = _659 + (((-_686.y) * _Globals.WhiteTint) * 0.0500000007450580596923828125);
|
||||
float _695 = _670 + ((_686.x * _Globals.WhiteTint) * 0.0500000007450580596923828125);
|
||||
float _700 = ((2.0 * _691) - (8.0 * _695)) + 4.0;
|
||||
float2 _706 = select(float2(_642, ((((-3.0) * _642) * _642) + (2.86999988555908203125 * _642)) - 0.2750000059604644775390625), _679, bool2(_Globals.WhiteTemp < 4000.0)) + (float2((3.0 * _691) / _700, (2.0 * _695) / _700) - _679);
|
||||
float2 _706 = select(float2(_642, (_642 * (((-3.0) * _642) + 2.86999988555908203125)) - 0.2750000059604644775390625), _679, bool2(_Globals.WhiteTemp < 4000.0)) + (float2((3.0 * _691) / _700, (2.0 * _695) / _700) - _679);
|
||||
float _707 = _706.x;
|
||||
float _708 = _706.y;
|
||||
float _709 = fast::max(_708, 1.0000000133514319600180897396058e-10);
|
||||
|
|
@ -327,7 +327,7 @@ fragment main0_out main0(main0_in in [[stage_in]], constant type_Globals& _Globa
|
|||
float3 _1215;
|
||||
if (_Globals.ColorShadow_Tint2.w == 0.0)
|
||||
{
|
||||
float3 _1157 = _3123;
|
||||
float3 _1157 = _3124;
|
||||
_1157.x = dot(_932, _Globals.ColorMatrixR_ColorCurveCd1.xyz);
|
||||
float3 _1162 = _1157;
|
||||
_1162.y = dot(_932, _Globals.ColorMatrixG_ColorCurveCd3Cm3.xyz);
|
||||
|
|
|
|||
|
|
@ -261,12 +261,6 @@ struct main0_in
|
|||
float4 in_var_TEXCOORD0_0 [[user(locn4)]];
|
||||
};
|
||||
|
||||
static inline __attribute__((always_inline))
|
||||
void _353()
|
||||
{
|
||||
discard_fragment();
|
||||
}
|
||||
|
||||
fragment main0_out main0(main0_in in [[stage_in]], constant type_View& View [[buffer(0)]], constant type_PrimitiveDither& PrimitiveDither [[buffer(1)]], constant type_PrimitiveFade& PrimitiveFade [[buffer(2)]], constant type_Material& Material [[buffer(3)]], texture2d<float> Material_Texture2D_0 [[texture(0)]], texture2d<float> Material_Texture2D_3 [[texture(1)]], sampler Material_Texture2D_0Sampler [[sampler(0)]], sampler Material_Texture2D_3Sampler [[sampler(1)]], float4 gl_FragCoord [[position]])
|
||||
{
|
||||
main0_out out = {};
|
||||
|
|
@ -340,13 +334,13 @@ fragment main0_out main0(main0_in in [[stage_in]], constant type_View& View [[bu
|
|||
float _317 = fract(cos(dot(floor(gl_FragCoord.xy), float2(347.834503173828125, 3343.28369140625))) * 1000.0);
|
||||
if ((float((PrimitiveDither.PrimitiveDither_LODFactor < 0.0) ? ((PrimitiveDither.PrimitiveDither_LODFactor + 1.0) > _317) : (PrimitiveDither.PrimitiveDither_LODFactor < _317)) - 0.001000000047497451305389404296875) < 0.0)
|
||||
{
|
||||
_353();
|
||||
discard_fragment();
|
||||
}
|
||||
}
|
||||
}
|
||||
if ((((_218.z + ((fast::min(fast::max(1.0 - (_218.x * Material.Material_ScalarExpressions[2].y), 0.0), 1.0) + ((_237 + (_242.x * Material.Material_ScalarExpressions[2].z)) * 0.16666667163372039794921875)) + (-0.5))) * ((fast::clamp((View.View_RealTime * PrimitiveFade.PrimitiveFade_FadeTimeScaleBias.x) + PrimitiveFade.PrimitiveFade_FadeTimeScaleBias.y, 0.0, 1.0) + ((_237 + _254.x) * 0.16666667163372039794921875)) + (-0.5))) - 0.33329999446868896484375) < 0.0)
|
||||
{
|
||||
_353();
|
||||
discard_fragment();
|
||||
}
|
||||
float2 _351 = ((((in.in_var_TEXCOORD6.xy / float2(in.in_var_TEXCOORD6.w)) - View.View_TemporalAAJitter.xy) - ((in.in_var_TEXCOORD7.xy / float2(in.in_var_TEXCOORD7.w)) - View.View_TemporalAAJitter.zw)) * float2(0.2495000064373016357421875)) + float2(0.49999237060546875);
|
||||
out.gl_FragDepth = fast::min(_140.z / (_140.w + (sqrt(dot(_272, _272)) / (fast::max(sqrt(dot(_276, _276)) / sqrt(dot(_279, _279)), sqrt(dot(_284, _284)) / sqrt(dot(_287, _287))) / abs(dot(float3x3(View.View_ViewToTranslatedWorld[0].xyz, View.View_ViewToTranslatedWorld[1].xyz, View.View_ViewToTranslatedWorld[2].xyz) * float3(0.0, 0.0, 1.0), _151))))), gl_FragCoord.z);
|
||||
|
|
|
|||
|
|
@ -263,7 +263,7 @@ struct type_StructuredBuffer_v4float
|
|||
float4 _m0[1];
|
||||
};
|
||||
|
||||
constant float4 _600 = {};
|
||||
constant float4 _602 = {};
|
||||
|
||||
struct main0_out
|
||||
{
|
||||
|
|
@ -328,7 +328,7 @@ kernel void main0(main0_in in [[stage_in]], constant type_View& View [[buffer(0)
|
|||
float4 _328 = float4(param_var_I[_313].FactoryInterpolants.InterpolantsVSToPS.TangentToWorld2.xyz, 0.0);
|
||||
float4 _336 = float4(param_var_I[_315].FactoryInterpolants.InterpolantsVSToPS.TangentToWorld2.xyz, 0.0);
|
||||
float4 _344 = float4(param_var_I[_316].FactoryInterpolants.InterpolantsVSToPS.TangentToWorld2.xyz, 0.0);
|
||||
spvUnsafeArray<float4, 3> _392 = spvUnsafeArray<float4, 3>({ param_var_I[gl_InvocationID].Position, (((((float4(2.0) * param_var_I[gl_InvocationID].Position) + param_var_I[_313].Position) - (float4(dot(param_var_I[_313].Position - param_var_I[gl_InvocationID].Position, _301)) * _301)) * float4(0.3333333432674407958984375)) + ((((float4(2.0) * param_var_I[_315].Position) + param_var_I[_316].Position) - (float4(dot(param_var_I[_316].Position - param_var_I[_315].Position, _336)) * _336)) * float4(0.3333333432674407958984375))) * float4(0.5), (((((float4(2.0) * param_var_I[_313].Position) + param_var_I[gl_InvocationID].Position) - (float4(dot(param_var_I[gl_InvocationID].Position - param_var_I[_313].Position, _328)) * _328)) * float4(0.3333333432674407958984375)) + ((((float4(2.0) * param_var_I[_316].Position) + param_var_I[_315].Position) - (float4(dot(param_var_I[_315].Position - param_var_I[_316].Position, _344)) * _344)) * float4(0.3333333432674407958984375))) * float4(0.5) });
|
||||
spvUnsafeArray<float4, 3> _392 = spvUnsafeArray<float4, 3>({ param_var_I[gl_InvocationID].Position, ((((float4(2.0) * param_var_I[gl_InvocationID].Position) + param_var_I[_313].Position) - (float4(dot(param_var_I[_313].Position - param_var_I[gl_InvocationID].Position, _301)) * _301)) + (((float4(2.0) * param_var_I[_315].Position) + param_var_I[_316].Position) - (float4(dot(param_var_I[_316].Position - param_var_I[_315].Position, _336)) * _336))) * float4(0.16666667163372039794921875), ((((float4(2.0) * param_var_I[_313].Position) + param_var_I[gl_InvocationID].Position) - (float4(dot(param_var_I[gl_InvocationID].Position - param_var_I[_313].Position, _328)) * _328)) + (((float4(2.0) * param_var_I[_316].Position) + param_var_I[_315].Position) - (float4(dot(param_var_I[_315].Position - param_var_I[_316].Position, _344)) * _344))) * float4(0.16666667163372039794921875) });
|
||||
gl_out[gl_InvocationID].out_var_TEXCOORD10_centroid = param_var_I[gl_InvocationID].FactoryInterpolants.InterpolantsVSToPS.TangentToWorld0;
|
||||
gl_out[gl_InvocationID].out_var_TEXCOORD11_centroid = param_var_I[gl_InvocationID].FactoryInterpolants.InterpolantsVSToPS.TangentToWorld2;
|
||||
gl_out[gl_InvocationID].out_var_COLOR0 = param_var_I[gl_InvocationID].FactoryInterpolants.InterpolantsVSToPS.Color;
|
||||
|
|
@ -346,7 +346,7 @@ kernel void main0(main0_in in [[stage_in]], constant type_View& View [[buffer(0)
|
|||
if (gl_InvocationID == 0u)
|
||||
{
|
||||
float4 _450 = (((((temp_var_hullMainRetVal[0u].WorldPosition[1] + temp_var_hullMainRetVal[0u].WorldPosition[2]) + temp_var_hullMainRetVal[1u].WorldPosition[1]) + temp_var_hullMainRetVal[1u].WorldPosition[2]) + temp_var_hullMainRetVal[2u].WorldPosition[1]) + temp_var_hullMainRetVal[2u].WorldPosition[2]) * float4(0.16666667163372039794921875);
|
||||
float4 _463 = _600;
|
||||
float4 _463 = _602;
|
||||
_463.x = 0.5 * (temp_var_hullMainRetVal[1u].TessellationMultiplier + temp_var_hullMainRetVal[2u].TessellationMultiplier);
|
||||
float4 _469 = _463;
|
||||
_469.y = 0.5 * (temp_var_hullMainRetVal[2u].TessellationMultiplier + temp_var_hullMainRetVal[0u].TessellationMultiplier);
|
||||
|
|
|
|||
|
|
@ -304,7 +304,7 @@ struct type_Primitive
|
|||
float4 Primitive_CustomPrimitiveData[4];
|
||||
};
|
||||
|
||||
constant float4 _611 = {};
|
||||
constant float4 _613 = {};
|
||||
|
||||
struct main0_out
|
||||
{
|
||||
|
|
@ -391,7 +391,7 @@ kernel void main0(main0_in in [[stage_in]], constant type_View& View [[buffer(0)
|
|||
float4 _336 = float4(param_var_I[_254].FactoryInterpolants.InterpolantsVSToPS.TangentToWorld2.xyz, 0.0);
|
||||
float4 _344 = float4(param_var_I[_256].FactoryInterpolants.InterpolantsVSToPS.TangentToWorld2.xyz, 0.0);
|
||||
float4 _352 = float4(param_var_I[_257].FactoryInterpolants.InterpolantsVSToPS.TangentToWorld2.xyz, 0.0);
|
||||
spvUnsafeArray<float4, 3> _402 = spvUnsafeArray<float4, 3>({ param_var_I[gl_InvocationID].Position, (((((float4(2.0) * param_var_I[gl_InvocationID].Position) + param_var_I[_254].Position) - (float4(dot(param_var_I[_254].Position - param_var_I[gl_InvocationID].Position, _247)) * _247)) * float4(0.3333333432674407958984375)) + ((((float4(2.0) * param_var_I[_256].Position) + param_var_I[_257].Position) - (float4(dot(param_var_I[_257].Position - param_var_I[_256].Position, _344)) * _344)) * float4(0.3333333432674407958984375))) * float4(0.5), (((((float4(2.0) * param_var_I[_254].Position) + param_var_I[gl_InvocationID].Position) - (float4(dot(param_var_I[gl_InvocationID].Position - param_var_I[_254].Position, _336)) * _336)) * float4(0.3333333432674407958984375)) + ((((float4(2.0) * param_var_I[_257].Position) + param_var_I[_256].Position) - (float4(dot(param_var_I[_256].Position - param_var_I[_257].Position, _352)) * _352)) * float4(0.3333333432674407958984375))) * float4(0.5) });
|
||||
spvUnsafeArray<float4, 3> _402 = spvUnsafeArray<float4, 3>({ param_var_I[gl_InvocationID].Position, ((((float4(2.0) * param_var_I[gl_InvocationID].Position) + param_var_I[_254].Position) - (float4(dot(param_var_I[_254].Position - param_var_I[gl_InvocationID].Position, _247)) * _247)) + (((float4(2.0) * param_var_I[_256].Position) + param_var_I[_257].Position) - (float4(dot(param_var_I[_257].Position - param_var_I[_256].Position, _344)) * _344))) * float4(0.16666667163372039794921875), ((((float4(2.0) * param_var_I[_254].Position) + param_var_I[gl_InvocationID].Position) - (float4(dot(param_var_I[gl_InvocationID].Position - param_var_I[_254].Position, _336)) * _336)) + (((float4(2.0) * param_var_I[_257].Position) + param_var_I[_256].Position) - (float4(dot(param_var_I[_256].Position - param_var_I[_257].Position, _352)) * _352))) * float4(0.16666667163372039794921875) });
|
||||
gl_out[gl_InvocationID].out_var_TEXCOORD10_centroid = param_var_I[gl_InvocationID].FactoryInterpolants.InterpolantsVSToPS.TangentToWorld0;
|
||||
gl_out[gl_InvocationID].out_var_TEXCOORD11_centroid = param_var_I[gl_InvocationID].FactoryInterpolants.InterpolantsVSToPS.TangentToWorld2;
|
||||
gl_out[gl_InvocationID].out_var_VS_To_DS_Position = param_var_I[gl_InvocationID].Position;
|
||||
|
|
@ -414,7 +414,7 @@ kernel void main0(main0_in in [[stage_in]], constant type_View& View [[buffer(0)
|
|||
if (gl_InvocationID == 0u)
|
||||
{
|
||||
float4 _461 = (((((temp_var_hullMainRetVal[0u].WorldPosition[1] + temp_var_hullMainRetVal[0u].WorldPosition[2]) + temp_var_hullMainRetVal[1u].WorldPosition[1]) + temp_var_hullMainRetVal[1u].WorldPosition[2]) + temp_var_hullMainRetVal[2u].WorldPosition[1]) + temp_var_hullMainRetVal[2u].WorldPosition[2]) * float4(0.16666667163372039794921875);
|
||||
float4 _474 = _611;
|
||||
float4 _474 = _613;
|
||||
_474.x = 0.5 * (temp_var_hullMainRetVal[1u].TessellationMultiplier + temp_var_hullMainRetVal[2u].TessellationMultiplier);
|
||||
float4 _480 = _474;
|
||||
_480.y = 0.5 * (temp_var_hullMainRetVal[2u].TessellationMultiplier + temp_var_hullMainRetVal[0u].TessellationMultiplier);
|
||||
|
|
|
|||
|
|
@ -286,7 +286,7 @@ struct type_Primitive
|
|||
float4 Primitive_CustomPrimitiveData[4];
|
||||
};
|
||||
|
||||
constant float4 _535 = {};
|
||||
constant float4 _537 = {};
|
||||
|
||||
struct main0_out
|
||||
{
|
||||
|
|
@ -343,7 +343,7 @@ kernel void main0(main0_in in [[stage_in]], constant type_View& View [[buffer(0)
|
|||
float4 _269 = float4(param_var_I[_254].FactoryInterpolants.InterpolantsVSToPS.TangentToWorld2.xyz, 0.0);
|
||||
float4 _277 = float4(param_var_I[_256].FactoryInterpolants.InterpolantsVSToPS.TangentToWorld2.xyz, 0.0);
|
||||
float4 _285 = float4(param_var_I[_257].FactoryInterpolants.InterpolantsVSToPS.TangentToWorld2.xyz, 0.0);
|
||||
spvUnsafeArray<float4, 3> _333 = spvUnsafeArray<float4, 3>({ param_var_I[gl_InvocationID].Position, (((((float4(2.0) * param_var_I[gl_InvocationID].Position) + param_var_I[_254].Position) - (float4(dot(param_var_I[_254].Position - param_var_I[gl_InvocationID].Position, _247)) * _247)) * float4(0.3333333432674407958984375)) + ((((float4(2.0) * param_var_I[_256].Position) + param_var_I[_257].Position) - (float4(dot(param_var_I[_257].Position - param_var_I[_256].Position, _277)) * _277)) * float4(0.3333333432674407958984375))) * float4(0.5), (((((float4(2.0) * param_var_I[_254].Position) + param_var_I[gl_InvocationID].Position) - (float4(dot(param_var_I[gl_InvocationID].Position - param_var_I[_254].Position, _269)) * _269)) * float4(0.3333333432674407958984375)) + ((((float4(2.0) * param_var_I[_257].Position) + param_var_I[_256].Position) - (float4(dot(param_var_I[_256].Position - param_var_I[_257].Position, _285)) * _285)) * float4(0.3333333432674407958984375))) * float4(0.5) });
|
||||
spvUnsafeArray<float4, 3> _333 = spvUnsafeArray<float4, 3>({ param_var_I[gl_InvocationID].Position, ((((float4(2.0) * param_var_I[gl_InvocationID].Position) + param_var_I[_254].Position) - (float4(dot(param_var_I[_254].Position - param_var_I[gl_InvocationID].Position, _247)) * _247)) + (((float4(2.0) * param_var_I[_256].Position) + param_var_I[_257].Position) - (float4(dot(param_var_I[_257].Position - param_var_I[_256].Position, _277)) * _277))) * float4(0.16666667163372039794921875), ((((float4(2.0) * param_var_I[_254].Position) + param_var_I[gl_InvocationID].Position) - (float4(dot(param_var_I[gl_InvocationID].Position - param_var_I[_254].Position, _269)) * _269)) + (((float4(2.0) * param_var_I[_257].Position) + param_var_I[_256].Position) - (float4(dot(param_var_I[_256].Position - param_var_I[_257].Position, _285)) * _285))) * float4(0.16666667163372039794921875) });
|
||||
gl_out[gl_InvocationID].out_var_TEXCOORD10_centroid = param_var_I[gl_InvocationID].FactoryInterpolants.InterpolantsVSToPS.TangentToWorld0;
|
||||
gl_out[gl_InvocationID].out_var_TEXCOORD11_centroid = param_var_I[gl_InvocationID].FactoryInterpolants.InterpolantsVSToPS.TangentToWorld2;
|
||||
gl_out[gl_InvocationID].out_var_COLOR0 = param_var_I[gl_InvocationID].FactoryInterpolants.InterpolantsVSToPS.Color;
|
||||
|
|
@ -358,7 +358,7 @@ kernel void main0(main0_in in [[stage_in]], constant type_View& View [[buffer(0)
|
|||
if (gl_InvocationID == 0u)
|
||||
{
|
||||
float4 _385 = (((((temp_var_hullMainRetVal[0u].WorldPosition[1] + temp_var_hullMainRetVal[0u].WorldPosition[2]) + temp_var_hullMainRetVal[1u].WorldPosition[1]) + temp_var_hullMainRetVal[1u].WorldPosition[2]) + temp_var_hullMainRetVal[2u].WorldPosition[1]) + temp_var_hullMainRetVal[2u].WorldPosition[2]) * float4(0.16666667163372039794921875);
|
||||
float4 _398 = _535;
|
||||
float4 _398 = _537;
|
||||
_398.x = 0.5 * (temp_var_hullMainRetVal[1u].TessellationMultiplier + temp_var_hullMainRetVal[2u].TessellationMultiplier);
|
||||
float4 _404 = _398;
|
||||
_404.y = 0.5 * (temp_var_hullMainRetVal[2u].TessellationMultiplier + temp_var_hullMainRetVal[0u].TessellationMultiplier);
|
||||
|
|
|
|||
|
|
@ -7,10 +7,9 @@ layout(location = 0) out vec4 b;
|
|||
void main()
|
||||
{
|
||||
vec4 _28 = (_ + a) + _;
|
||||
vec4 _34 = (_ - a) + a;
|
||||
b = _28;
|
||||
b = _34;
|
||||
b = _;
|
||||
b = _28;
|
||||
b = _34;
|
||||
b = _;
|
||||
}
|
||||
|
||||
|
|
|
|||
|
|
@ -1,32 +0,0 @@
|
|||
#ifndef SPIRV_CROSS_CONSTANT_ID_0
|
||||
#define SPIRV_CROSS_CONSTANT_ID_0 1u
|
||||
#endif
|
||||
static const uint _3 = SPIRV_CROSS_CONSTANT_ID_0;
|
||||
#ifndef SPIRV_CROSS_CONSTANT_ID_2
|
||||
#define SPIRV_CROSS_CONSTANT_ID_2 3u
|
||||
#endif
|
||||
static const uint _4 = SPIRV_CROSS_CONSTANT_ID_2;
|
||||
static const uint3 gl_WorkGroupSize = uint3(_3, 2u, _4);
|
||||
|
||||
RWByteAddressBuffer _8 : register(u0);
|
||||
RWByteAddressBuffer _9 : register(u1);
|
||||
|
||||
static uint3 gl_WorkGroupID;
|
||||
struct SPIRV_Cross_Input
|
||||
{
|
||||
uint3 gl_WorkGroupID : SV_GroupID;
|
||||
};
|
||||
|
||||
static uint3 _22 = gl_WorkGroupSize;
|
||||
|
||||
void comp_main()
|
||||
{
|
||||
_8.Store(gl_WorkGroupID.x * 4 + 0, asuint(asfloat(_9.Load(gl_WorkGroupID.x * 4 + 0)) + asfloat(_8.Load(gl_WorkGroupID.x * 4 + 0))));
|
||||
}
|
||||
|
||||
[numthreads(SPIRV_CROSS_CONSTANT_ID_0, 2, SPIRV_CROSS_CONSTANT_ID_2)]
|
||||
void main(SPIRV_Cross_Input stage_input)
|
||||
{
|
||||
gl_WorkGroupID = stage_input.gl_WorkGroupID;
|
||||
comp_main();
|
||||
}
|
||||
|
|
@ -1,44 +0,0 @@
|
|||
struct CBO_1
|
||||
{
|
||||
float4 a;
|
||||
float4 b;
|
||||
float4 c;
|
||||
float4 d;
|
||||
};
|
||||
|
||||
ConstantBuffer<CBO_1> cbo[2][4] : register(b4, space0);
|
||||
cbuffer PushMe
|
||||
{
|
||||
float4 push_a : packoffset(c0);
|
||||
float4 push_b : packoffset(c1);
|
||||
float4 push_c : packoffset(c2);
|
||||
float4 push_d : packoffset(c3);
|
||||
};
|
||||
|
||||
|
||||
static float4 FragColor;
|
||||
|
||||
struct SPIRV_Cross_Output
|
||||
{
|
||||
float4 FragColor : SV_Target0;
|
||||
};
|
||||
|
||||
void frag_main()
|
||||
{
|
||||
FragColor = cbo[1][2].a;
|
||||
FragColor += cbo[1][2].b;
|
||||
FragColor += cbo[1][2].c;
|
||||
FragColor += cbo[1][2].d;
|
||||
FragColor += push_a;
|
||||
FragColor += push_b;
|
||||
FragColor += push_c;
|
||||
FragColor += push_d;
|
||||
}
|
||||
|
||||
SPIRV_Cross_Output main()
|
||||
{
|
||||
frag_main();
|
||||
SPIRV_Cross_Output stage_output;
|
||||
stage_output.FragColor = FragColor;
|
||||
return stage_output;
|
||||
}
|
||||
|
|
@ -16,7 +16,7 @@ struct SSBO2
|
|||
|
||||
constant uint3 gl_WorkGroupSize [[maybe_unused]] = uint3(1u);
|
||||
|
||||
kernel void main0(const device SSBO& _24 [[buffer(0)]], device SSBO2& _177 [[buffer(1)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
|
||||
kernel void main0(const device SSBO& _24 [[buffer(0)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
|
||||
{
|
||||
uint ident = gl_GlobalInvocationID.x;
|
||||
float4 idat = _24.in_data[ident];
|
||||
|
|
@ -87,25 +87,5 @@ kernel void main0(const device SSBO& _24 [[buffer(0)]], device SSBO2& _177 [[buf
|
|||
k += 10;
|
||||
continue;
|
||||
}
|
||||
k = 0;
|
||||
do
|
||||
{
|
||||
k++;
|
||||
} while (k > 10);
|
||||
int l = 0;
|
||||
for (;;)
|
||||
{
|
||||
if (l == 5)
|
||||
{
|
||||
int _172 = l;
|
||||
l = _172 + 1;
|
||||
continue;
|
||||
}
|
||||
idat += float4(1.0);
|
||||
int _172 = l;
|
||||
l = _172 + 1;
|
||||
continue;
|
||||
}
|
||||
_177.out_data[ident] = idat;
|
||||
}
|
||||
|
||||
|
|
|
|||
|
|
@ -25,7 +25,8 @@ kernel void main0(device SSBO2& _27 [[buffer(0)]], uint3 gl_GlobalInvocationID [
|
|||
return;
|
||||
}
|
||||
}
|
||||
for (int i = 0; i < 20; i++)
|
||||
int i = 0;
|
||||
while (i < 20)
|
||||
{
|
||||
if (i == 10)
|
||||
{
|
||||
|
|
|
|||
76
3rdparty/spirv-cross/reference/shaders-msl-no-opt/packing/load-store-col-rows.comp
vendored
Normal file
76
3rdparty/spirv-cross/reference/shaders-msl-no-opt/packing/load-store-col-rows.comp
vendored
Normal file
|
|
@ -0,0 +1,76 @@
|
|||
#pragma clang diagnostic ignored "-Wmissing-prototypes"
|
||||
|
||||
#include <metal_stdlib>
|
||||
#include <simd/simd.h>
|
||||
|
||||
using namespace metal;
|
||||
|
||||
typedef packed_float3 packed_float2x3[2];
|
||||
typedef packed_float3 packed_rm_float3x2[2];
|
||||
|
||||
struct SSBO1
|
||||
{
|
||||
float2x4 a;
|
||||
float2x4 a2;
|
||||
};
|
||||
|
||||
struct SSBO2
|
||||
{
|
||||
packed_float2x3 b;
|
||||
packed_rm_float3x2 b2;
|
||||
};
|
||||
|
||||
constant uint3 gl_WorkGroupSize [[maybe_unused]] = uint3(1u);
|
||||
|
||||
static inline __attribute__((always_inline))
|
||||
void load_store_column(device SSBO1& v_21)
|
||||
{
|
||||
float2 u = v_21.a[0].xy;
|
||||
float2 v = v_21.a[1].xy;
|
||||
u += v;
|
||||
v_21.a[0].xy = u;
|
||||
v_21.a[1].xy = v;
|
||||
}
|
||||
|
||||
static inline __attribute__((always_inline))
|
||||
void load_store_row(device SSBO1& v_21)
|
||||
{
|
||||
float2 u = float2(v_21.a2[0][0], v_21.a2[1][0]);
|
||||
float2 v = float2(v_21.a2[0][1], v_21.a2[1][1]);
|
||||
u += v;
|
||||
v_21.a2[0][0] = u.x;
|
||||
v_21.a2[1][0] = u.y;
|
||||
v_21.a2[0][1] = v.x;
|
||||
v_21.a2[1][1] = v.y;
|
||||
}
|
||||
|
||||
static inline __attribute__((always_inline))
|
||||
void load_store_packed_column(device SSBO2& v_58)
|
||||
{
|
||||
float3 u = float3(v_58.b[0]);
|
||||
float3 v = float3(v_58.b[1]);
|
||||
u += v;
|
||||
v_58.b[0] = u;
|
||||
v_58.b[1] = v;
|
||||
}
|
||||
|
||||
static inline __attribute__((always_inline))
|
||||
void load_store_packed_row(device SSBO2& v_58)
|
||||
{
|
||||
float2 u = float2(v_58.b2[0][0], v_58.b2[1][0]);
|
||||
float2 v = float2(v_58.b2[0][1], v_58.b2[1][1]);
|
||||
u += v;
|
||||
v_58.b2[0][0] = u.x;
|
||||
v_58.b2[1][0] = u.y;
|
||||
v_58.b2[0][1] = v.x;
|
||||
v_58.b2[1][1] = v.y;
|
||||
}
|
||||
|
||||
kernel void main0(device SSBO1& v_21 [[buffer(0)]], device SSBO2& v_58 [[buffer(1)]])
|
||||
{
|
||||
load_store_column(v_21);
|
||||
load_store_row(v_21);
|
||||
load_store_packed_column(v_58);
|
||||
load_store_packed_row(v_58);
|
||||
}
|
||||
|
||||
|
|
@ -27,8 +27,6 @@ struct VertexConstantsBuffer
|
|||
float4 vc[16];
|
||||
};
|
||||
|
||||
constant float4 _295 = {};
|
||||
|
||||
struct main0_out
|
||||
{
|
||||
float4 tc0 [[user(locn0)]];
|
||||
|
|
|
|||
|
|
@ -1,22 +0,0 @@
|
|||
#include <metal_stdlib>
|
||||
#include <simd/simd.h>
|
||||
|
||||
using namespace metal;
|
||||
|
||||
struct _6
|
||||
{
|
||||
float _m0[1];
|
||||
};
|
||||
|
||||
constant uint _3_tmp [[function_constant(0)]];
|
||||
constant uint _3 = is_function_constant_defined(_3_tmp) ? _3_tmp : 1u;
|
||||
constant uint _4_tmp [[function_constant(2)]];
|
||||
constant uint _4 = is_function_constant_defined(_4_tmp) ? _4_tmp : 3u;
|
||||
constant uint3 gl_WorkGroupSize [[maybe_unused]] = uint3(_3, 2u, _4);
|
||||
|
||||
kernel void main0(device _6& _8 [[buffer(0)]], device _6& _9 [[buffer(1)]], uint3 gl_WorkGroupID [[threadgroup_position_in_grid]])
|
||||
{
|
||||
_8._m0[gl_WorkGroupID.x] = _9._m0[gl_WorkGroupID.x] + _8._m0[gl_WorkGroupID.x];
|
||||
uint3 _23 = gl_WorkGroupSize;
|
||||
}
|
||||
|
||||
|
|
@ -1,21 +0,0 @@
|
|||
#include <metal_stdlib>
|
||||
#include <simd/simd.h>
|
||||
|
||||
using namespace metal;
|
||||
|
||||
struct main0_out
|
||||
{
|
||||
half foo [[color(0)]];
|
||||
short bar [[color(1)]];
|
||||
ushort baz [[color(2)]];
|
||||
};
|
||||
|
||||
fragment main0_out main0()
|
||||
{
|
||||
main0_out out = {};
|
||||
out.foo = half(1.0);
|
||||
out.bar = short(2);
|
||||
out.baz = ushort(3);
|
||||
return out;
|
||||
}
|
||||
|
||||
|
|
@ -1,24 +0,0 @@
|
|||
#include <metal_stdlib>
|
||||
#include <simd/simd.h>
|
||||
|
||||
using namespace metal;
|
||||
|
||||
struct main0_out
|
||||
{
|
||||
float4 gl_Position [[position]];
|
||||
uint gl_Layer [[render_target_array_index]];
|
||||
};
|
||||
|
||||
struct main0_in
|
||||
{
|
||||
float4 coord [[attribute(0)]];
|
||||
};
|
||||
|
||||
vertex main0_out main0(main0_in in [[stage_in]])
|
||||
{
|
||||
main0_out out = {};
|
||||
out.gl_Position = in.coord;
|
||||
out.gl_Layer = uint(int(in.coord.z));
|
||||
return out;
|
||||
}
|
||||
|
||||
|
|
@ -40,11 +40,11 @@ vertex main0_out main0(main0_in in [[stage_in]], constant _15& _17 [[buffer(0)]]
|
|||
main0_out out = {};
|
||||
float3 _91;
|
||||
float3 _13;
|
||||
do
|
||||
for (;;)
|
||||
{
|
||||
_13 = normalize(float4(in.m_25.xyz, 0.0) * _17._m1);
|
||||
break;
|
||||
} while (false);
|
||||
}
|
||||
float4 _39 = _44._m0 * float4(float3(_44._m3) + (in.m_25.xyz * (_44._m6 + _44._m7)), 1.0);
|
||||
out.m_72 = _13;
|
||||
float4 _74 = _39;
|
||||
|
|
|
|||
|
|
@ -1,24 +0,0 @@
|
|||
#include <metal_stdlib>
|
||||
#include <simd/simd.h>
|
||||
|
||||
using namespace metal;
|
||||
|
||||
struct main0_out
|
||||
{
|
||||
float4 gl_Position [[position]];
|
||||
uint gl_ViewportIndex [[viewport_array_index]];
|
||||
};
|
||||
|
||||
struct main0_in
|
||||
{
|
||||
float4 coord [[attribute(0)]];
|
||||
};
|
||||
|
||||
vertex main0_out main0(main0_in in [[stage_in]])
|
||||
{
|
||||
main0_out out = {};
|
||||
out.gl_Position = in.coord;
|
||||
out.gl_ViewportIndex = uint(int(in.coord.z));
|
||||
return out;
|
||||
}
|
||||
|
||||
|
|
@ -1,150 +0,0 @@
|
|||
#include <metal_stdlib>
|
||||
#include <simd/simd.h>
|
||||
|
||||
using namespace metal;
|
||||
|
||||
typedef packed_float2 packed_float2x2[2];
|
||||
typedef packed_float3 packed_float2x3[2];
|
||||
typedef packed_float3 packed_rm_float3x2[2];
|
||||
|
||||
struct S0
|
||||
{
|
||||
packed_float2 a[1];
|
||||
float b;
|
||||
};
|
||||
|
||||
struct S1
|
||||
{
|
||||
packed_float3 a;
|
||||
float b;
|
||||
};
|
||||
|
||||
struct S2
|
||||
{
|
||||
packed_float3 a[1];
|
||||
float b;
|
||||
};
|
||||
|
||||
struct S3
|
||||
{
|
||||
packed_float2 a;
|
||||
float b;
|
||||
};
|
||||
|
||||
struct Content
|
||||
{
|
||||
S0 m0s[1];
|
||||
S1 m1s[1];
|
||||
S2 m2s[1];
|
||||
S0 m0;
|
||||
S1 m1;
|
||||
S2 m2;
|
||||
S3 m3;
|
||||
float m4;
|
||||
};
|
||||
|
||||
struct SSBO1
|
||||
{
|
||||
Content content;
|
||||
Content content1[2];
|
||||
Content content2;
|
||||
float2x2 m0;
|
||||
float2x2 m1;
|
||||
packed_float2x3 m2[4];
|
||||
float3x2 m3;
|
||||
float2x2 m4;
|
||||
float2x2 m5[9];
|
||||
float3x2 m6[4][2];
|
||||
packed_rm_float3x2 m7;
|
||||
float array[1];
|
||||
};
|
||||
|
||||
struct S0_1
|
||||
{
|
||||
float4 a[1];
|
||||
float b;
|
||||
char _m0_final_padding[12];
|
||||
};
|
||||
|
||||
struct S1_1
|
||||
{
|
||||
packed_float3 a;
|
||||
float b;
|
||||
};
|
||||
|
||||
struct S2_1
|
||||
{
|
||||
float3 a[1];
|
||||
float b;
|
||||
char _m0_final_padding[12];
|
||||
};
|
||||
|
||||
struct S3_1
|
||||
{
|
||||
float2 a;
|
||||
float b;
|
||||
};
|
||||
|
||||
struct Content_1
|
||||
{
|
||||
S0_1 m0s[1];
|
||||
S1_1 m1s[1];
|
||||
S2_1 m2s[1];
|
||||
S0_1 m0;
|
||||
S1_1 m1;
|
||||
S2_1 m2;
|
||||
S3_1 m3;
|
||||
float m4;
|
||||
char _m0_final_padding[12];
|
||||
};
|
||||
|
||||
struct SSBO0
|
||||
{
|
||||
Content_1 content;
|
||||
Content_1 content1[2];
|
||||
Content_1 content2;
|
||||
float2x4 m0;
|
||||
float2x4 m1;
|
||||
float2x3 m2[4];
|
||||
float3x4 m3;
|
||||
float2x4 m4;
|
||||
float2x4 m5[9];
|
||||
float3x4 m6[4][2];
|
||||
float2x3 m7;
|
||||
float4 array[1];
|
||||
};
|
||||
|
||||
struct SSBO2
|
||||
{
|
||||
float m0;
|
||||
packed_float2x2 m1;
|
||||
packed_rm_float3x2 m2;
|
||||
};
|
||||
|
||||
constant uint3 gl_WorkGroupSize [[maybe_unused]] = uint3(1u);
|
||||
|
||||
kernel void main0(device SSBO1& ssbo_scalar [[buffer(0)]], device SSBO0& ssbo_140 [[buffer(1)]], device SSBO2& ssbo_scalar2 [[buffer(2)]])
|
||||
{
|
||||
ssbo_scalar.content.m0s[0].a[0] = ssbo_140.content.m0s[0].a[0].xy;
|
||||
ssbo_scalar.content.m0s[0].b = ssbo_140.content.m0s[0].b;
|
||||
ssbo_scalar.content.m1s[0].a = float3(ssbo_140.content.m1s[0].a);
|
||||
ssbo_scalar.content.m1s[0].b = ssbo_140.content.m1s[0].b;
|
||||
ssbo_scalar.content.m2s[0].a[0] = ssbo_140.content.m2s[0].a[0];
|
||||
ssbo_scalar.content.m2s[0].b = ssbo_140.content.m2s[0].b;
|
||||
ssbo_scalar.content.m0.a[0] = ssbo_140.content.m0.a[0].xy;
|
||||
ssbo_scalar.content.m0.b = ssbo_140.content.m0.b;
|
||||
ssbo_scalar.content.m1.a = float3(ssbo_140.content.m1.a);
|
||||
ssbo_scalar.content.m1.b = ssbo_140.content.m1.b;
|
||||
ssbo_scalar.content.m2.a[0] = ssbo_140.content.m2.a[0];
|
||||
ssbo_scalar.content.m2.b = ssbo_140.content.m2.b;
|
||||
ssbo_scalar.content.m3.a = ssbo_140.content.m3.a;
|
||||
ssbo_scalar.content.m3.b = ssbo_140.content.m3.b;
|
||||
ssbo_scalar.content.m4 = ssbo_140.content.m4;
|
||||
ssbo_scalar.content.m1.a = float2x3(float3(ssbo_scalar.m2[1][0]), float3(ssbo_scalar.m2[1][1])) * float2(ssbo_scalar.content.m0.a[0]);
|
||||
ssbo_scalar.m0 = float2x2(float2(ssbo_scalar2.m1[0]), float2(ssbo_scalar2.m1[1]));
|
||||
ssbo_scalar2.m1[0] = float2(ssbo_scalar.m4[0][0], ssbo_scalar.m4[1][0]);
|
||||
ssbo_scalar2.m1[1] = float2(ssbo_scalar.m4[0][1], ssbo_scalar.m4[1][1]);
|
||||
ssbo_scalar2.m2[0] = float3(ssbo_scalar.m3[0][0], ssbo_scalar.m3[1][0], ssbo_scalar.m3[2][0]);
|
||||
ssbo_scalar2.m2[1] = float3(ssbo_scalar.m3[0][1], ssbo_scalar.m3[1][1], ssbo_scalar.m3[2][1]);
|
||||
}
|
||||
|
||||
|
|
@ -1,11 +0,0 @@
|
|||
#include <metal_stdlib>
|
||||
#include <simd/simd.h>
|
||||
|
||||
using namespace metal;
|
||||
|
||||
fragment void main0()
|
||||
{
|
||||
bool _9 = simd_is_helper_thread();
|
||||
bool helper = _9;
|
||||
}
|
||||
|
||||
|
|
@ -1,36 +0,0 @@
|
|||
#include <metal_stdlib>
|
||||
#include <simd/simd.h>
|
||||
|
||||
using namespace metal;
|
||||
|
||||
struct UBO
|
||||
{
|
||||
float a[1];
|
||||
float2 b[2];
|
||||
};
|
||||
|
||||
struct UBOEnhancedLayout
|
||||
{
|
||||
float c[1];
|
||||
float2 d[2];
|
||||
char _m2_pad[9976];
|
||||
float e;
|
||||
};
|
||||
|
||||
struct main0_out
|
||||
{
|
||||
float FragColor [[color(0)]];
|
||||
};
|
||||
|
||||
struct main0_in
|
||||
{
|
||||
int vIndex [[user(locn0)]];
|
||||
};
|
||||
|
||||
fragment main0_out main0(main0_in in [[stage_in]], constant UBO& _17 [[buffer(0)]], constant UBOEnhancedLayout& _30 [[buffer(1)]])
|
||||
{
|
||||
main0_out out = {};
|
||||
out.FragColor = (_17.a[in.vIndex] + _30.c[in.vIndex]) + _30.e;
|
||||
return out;
|
||||
}
|
||||
|
||||
|
|
@ -7,11 +7,6 @@ layout(binding = 0, std430) readonly buffer SSBO
|
|||
vec4 in_data[];
|
||||
} _24;
|
||||
|
||||
layout(binding = 1, std430) writeonly buffer SSBO2
|
||||
{
|
||||
vec4 out_data[];
|
||||
} _177;
|
||||
|
||||
void main()
|
||||
{
|
||||
uint ident = gl_GlobalInvocationID.x;
|
||||
|
|
@ -83,25 +78,5 @@ void main()
|
|||
k += 10;
|
||||
continue;
|
||||
}
|
||||
k = 0;
|
||||
do
|
||||
{
|
||||
k++;
|
||||
} while (k > 10);
|
||||
int l = 0;
|
||||
for (;;)
|
||||
{
|
||||
if (l == 5)
|
||||
{
|
||||
int _172 = l;
|
||||
l = _172 + 1;
|
||||
continue;
|
||||
}
|
||||
idat += vec4(1.0);
|
||||
int _172 = l;
|
||||
l = _172 + 1;
|
||||
continue;
|
||||
}
|
||||
_177.out_data[ident] = idat;
|
||||
}
|
||||
|
||||
|
|
|
|||
|
|
@ -21,7 +21,8 @@ void main()
|
|||
return;
|
||||
}
|
||||
}
|
||||
for (int i = 0; i < 20; i++)
|
||||
int i = 0;
|
||||
while (i < 20)
|
||||
{
|
||||
if (i == 10)
|
||||
{
|
||||
|
|
|
|||
|
|
@ -1,11 +0,0 @@
|
|||
#version 450
|
||||
#extension GL_AMD_shader_fragment_mask : require
|
||||
|
||||
layout(input_attachment_index = 0, set = 0, binding = 0) uniform subpassInputMS t;
|
||||
|
||||
void main()
|
||||
{
|
||||
vec4 test2 = fragmentFetchAMD(t, 4u);
|
||||
uint testi2 = fragmentMaskFetchAMD(t);
|
||||
}
|
||||
|
||||
|
|
@ -1,15 +0,0 @@
|
|||
#version 450
|
||||
#extension GL_AMD_shader_fragment_mask : require
|
||||
#extension GL_AMD_shader_explicit_vertex_parameter : require
|
||||
|
||||
layout(binding = 0) uniform sampler2DMS texture1;
|
||||
|
||||
layout(location = 0) __explicitInterpAMD in vec4 vary;
|
||||
|
||||
void main()
|
||||
{
|
||||
uint testi1 = fragmentMaskFetchAMD(texture1, ivec2(0));
|
||||
vec4 test1 = fragmentFetchAMD(texture1, ivec2(1), 2u);
|
||||
vec4 pos = interpolateAtVertexAMD(vary, 0u);
|
||||
}
|
||||
|
||||
|
|
@ -1,11 +0,0 @@
|
|||
#version 450
|
||||
#extension GL_AMD_shader_ballot : require
|
||||
layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
|
||||
|
||||
void main()
|
||||
{
|
||||
float addInvocations = addInvocationsNonUniformAMD(0.0);
|
||||
int minInvocations = minInvocationsNonUniformAMD(1);
|
||||
uint maxInvocations = uint(maxInvocationsNonUniformAMD(4));
|
||||
}
|
||||
|
||||
|
|
@ -1,28 +0,0 @@
|
|||
#version 450
|
||||
|
||||
#ifndef SPIRV_CROSS_CONSTANT_ID_0
|
||||
#define SPIRV_CROSS_CONSTANT_ID_0 1u
|
||||
#endif
|
||||
#ifndef SPIRV_CROSS_CONSTANT_ID_2
|
||||
#define SPIRV_CROSS_CONSTANT_ID_2 3u
|
||||
#endif
|
||||
|
||||
layout(local_size_x = SPIRV_CROSS_CONSTANT_ID_0, local_size_y = 2, local_size_z = SPIRV_CROSS_CONSTANT_ID_2) in;
|
||||
|
||||
layout(binding = 0, std430) buffer _6_8
|
||||
{
|
||||
float _m0[];
|
||||
} _8;
|
||||
|
||||
layout(binding = 1, std430) buffer _6_9
|
||||
{
|
||||
float _m0[];
|
||||
} _9;
|
||||
|
||||
uvec3 _22 = gl_WorkGroupSize;
|
||||
|
||||
void main()
|
||||
{
|
||||
_8._m0[gl_WorkGroupID.x] = _9._m0[gl_WorkGroupID.x] + _8._m0[gl_WorkGroupID.x];
|
||||
}
|
||||
|
||||
|
|
@ -1,18 +0,0 @@
|
|||
#version 430
|
||||
layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
|
||||
|
||||
layout(binding = 0, std430) buffer _8_9
|
||||
{
|
||||
float _m0[];
|
||||
} _9;
|
||||
|
||||
layout(binding = 1, std430) buffer _8_10
|
||||
{
|
||||
float _m0[];
|
||||
} _10;
|
||||
|
||||
void main()
|
||||
{
|
||||
_10._m0[gl_GlobalInvocationID.x] = -_9._m0[gl_GlobalInvocationID.x];
|
||||
}
|
||||
|
||||
|
|
@ -1,41 +0,0 @@
|
|||
#version 450
|
||||
layout(triangles) in;
|
||||
layout(max_vertices = 3, triangle_strip) out;
|
||||
|
||||
struct VertexOutput
|
||||
{
|
||||
vec4 pos;
|
||||
};
|
||||
|
||||
struct GeometryOutput
|
||||
{
|
||||
vec4 pos;
|
||||
uint layer;
|
||||
};
|
||||
|
||||
void _main(VertexOutput _input[3], GeometryOutput stream)
|
||||
{
|
||||
GeometryOutput _output;
|
||||
_output.layer = 1u;
|
||||
for (int v = 0; v < 3; v++)
|
||||
{
|
||||
_output.pos = _input[v].pos;
|
||||
gl_Position = _output.pos;
|
||||
gl_Layer = int(_output.layer);
|
||||
EmitVertex();
|
||||
}
|
||||
EndPrimitive();
|
||||
}
|
||||
|
||||
void main()
|
||||
{
|
||||
VertexOutput _input[3];
|
||||
_input[0].pos = gl_in[0].gl_Position;
|
||||
_input[1].pos = gl_in[1].gl_Position;
|
||||
_input[2].pos = gl_in[2].gl_Position;
|
||||
VertexOutput param[3] = _input;
|
||||
GeometryOutput param_1;
|
||||
_main(param, param_1);
|
||||
GeometryOutput stream = param_1;
|
||||
}
|
||||
|
||||
|
|
@ -1,79 +0,0 @@
|
|||
#version 450
|
||||
layout(vertices = 3) out;
|
||||
|
||||
struct VertexOutput
|
||||
{
|
||||
vec4 pos;
|
||||
vec2 uv;
|
||||
};
|
||||
|
||||
struct HSOut
|
||||
{
|
||||
vec4 pos;
|
||||
vec2 uv;
|
||||
};
|
||||
|
||||
struct HSConstantOut
|
||||
{
|
||||
float EdgeTess[3];
|
||||
float InsideTess;
|
||||
};
|
||||
|
||||
struct VertexOutput_1
|
||||
{
|
||||
vec2 uv;
|
||||
};
|
||||
|
||||
struct HSOut_1
|
||||
{
|
||||
vec2 uv;
|
||||
};
|
||||
|
||||
layout(location = 0) in VertexOutput_1 p[];
|
||||
layout(location = 0) out HSOut_1 _entryPointOutput[3];
|
||||
|
||||
HSOut _hs_main(VertexOutput p_1[3], uint i)
|
||||
{
|
||||
HSOut _output;
|
||||
_output.pos = p_1[i].pos;
|
||||
_output.uv = p_1[i].uv;
|
||||
return _output;
|
||||
}
|
||||
|
||||
HSConstantOut PatchHS(VertexOutput _patch[3])
|
||||
{
|
||||
HSConstantOut _output;
|
||||
_output.EdgeTess[0] = (vec2(1.0) + _patch[0].uv).x;
|
||||
_output.EdgeTess[1] = (vec2(1.0) + _patch[0].uv).x;
|
||||
_output.EdgeTess[2] = (vec2(1.0) + _patch[0].uv).x;
|
||||
_output.InsideTess = (vec2(1.0) + _patch[0].uv).x;
|
||||
return _output;
|
||||
}
|
||||
|
||||
void main()
|
||||
{
|
||||
VertexOutput p_1[3];
|
||||
p_1[0].pos = gl_in[0].gl_Position;
|
||||
p_1[0].uv = p[0].uv;
|
||||
p_1[1].pos = gl_in[1].gl_Position;
|
||||
p_1[1].uv = p[1].uv;
|
||||
p_1[2].pos = gl_in[2].gl_Position;
|
||||
p_1[2].uv = p[2].uv;
|
||||
uint i = gl_InvocationID;
|
||||
VertexOutput param[3] = p_1;
|
||||
uint param_1 = i;
|
||||
HSOut flattenTemp = _hs_main(param, param_1);
|
||||
gl_out[gl_InvocationID].gl_Position = flattenTemp.pos;
|
||||
_entryPointOutput[gl_InvocationID].uv = flattenTemp.uv;
|
||||
barrier();
|
||||
if (int(gl_InvocationID) == 0)
|
||||
{
|
||||
VertexOutput param_2[3] = p_1;
|
||||
HSConstantOut _patchConstantResult = PatchHS(param_2);
|
||||
gl_TessLevelOuter[0] = _patchConstantResult.EdgeTess[0];
|
||||
gl_TessLevelOuter[1] = _patchConstantResult.EdgeTess[1];
|
||||
gl_TessLevelOuter[2] = _patchConstantResult.EdgeTess[2];
|
||||
gl_TessLevelInner[0] = _patchConstantResult.InsideTess;
|
||||
}
|
||||
}
|
||||
|
||||
|
|
@ -1,34 +0,0 @@
|
|||
#version 450
|
||||
#if defined(GL_AMD_gpu_shader_half_float)
|
||||
#extension GL_AMD_gpu_shader_half_float : require
|
||||
#elif defined(GL_NV_gpu_shader5)
|
||||
#extension GL_NV_gpu_shader5 : require
|
||||
#else
|
||||
#error No extension available for FP16.
|
||||
#endif
|
||||
#if defined(GL_AMD_gpu_shader_int16)
|
||||
#extension GL_AMD_gpu_shader_int16 : require
|
||||
#else
|
||||
#error No extension available for Int16.
|
||||
#endif
|
||||
layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
|
||||
|
||||
layout(binding = 0, std430) buffer SSBO0
|
||||
{
|
||||
i16vec4 inputs[];
|
||||
} _25;
|
||||
|
||||
layout(binding = 1, std430) buffer SSBO1
|
||||
{
|
||||
ivec4 outputs[];
|
||||
} _39;
|
||||
|
||||
void main()
|
||||
{
|
||||
uint ident = gl_GlobalInvocationID.x;
|
||||
f16vec2 a = int16BitsToFloat16(_25.inputs[ident].xy);
|
||||
_39.outputs[ident].x = int(packFloat2x16(a + f16vec2(float16_t(1.0))));
|
||||
_39.outputs[ident].y = packInt2x16(_25.inputs[ident].zw);
|
||||
_39.outputs[ident].z = int(packUint2x16(u16vec2(_25.inputs[ident].xy)));
|
||||
}
|
||||
|
||||
|
|
@ -1,39 +0,0 @@
|
|||
#version 450
|
||||
#if defined(GL_AMD_gpu_shader_int16)
|
||||
#extension GL_AMD_gpu_shader_int16 : require
|
||||
#else
|
||||
#error No extension available for Int16.
|
||||
#endif
|
||||
#if defined(GL_AMD_gpu_shader_half_float)
|
||||
#extension GL_AMD_gpu_shader_half_float : require
|
||||
#elif defined(GL_NV_gpu_shader5)
|
||||
#extension GL_NV_gpu_shader5 : require
|
||||
#else
|
||||
#error No extension available for FP16.
|
||||
#endif
|
||||
layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
|
||||
|
||||
layout(binding = 1, std430) buffer SSBO1
|
||||
{
|
||||
i16vec4 outputs[];
|
||||
} _21;
|
||||
|
||||
layout(binding = 0, std430) buffer SSBO0
|
||||
{
|
||||
ivec4 inputs[];
|
||||
} _29;
|
||||
|
||||
layout(binding = 2, std140) uniform UBO
|
||||
{
|
||||
f16vec4 const0;
|
||||
} _40;
|
||||
|
||||
void main()
|
||||
{
|
||||
uint ident = gl_GlobalInvocationID.x;
|
||||
i16vec2 _47 = unpackInt2x16(_29.inputs[ident].x) + float16BitsToInt16(_40.const0.xy);
|
||||
_21.outputs[ident] = i16vec4(_47.x, _47.y, _21.outputs[ident].z, _21.outputs[ident].w);
|
||||
i16vec2 _66 = i16vec2(unpackUint2x16(uint(_29.inputs[ident].y)) - float16BitsToUint16(_40.const0.zw));
|
||||
_21.outputs[ident] = i16vec4(_21.outputs[ident].x, _21.outputs[ident].y, _66.x, _66.y);
|
||||
}
|
||||
|
||||
|
|
@ -1,65 +0,0 @@
|
|||
#version 310 es
|
||||
layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
|
||||
|
||||
struct Foo
|
||||
{
|
||||
vec4 a;
|
||||
vec4 b;
|
||||
vec4 c;
|
||||
vec4 d;
|
||||
};
|
||||
|
||||
layout(binding = 1, std430) readonly buffer SSBO2
|
||||
{
|
||||
vec4 data[];
|
||||
} indata;
|
||||
|
||||
layout(binding = 0, std430) writeonly buffer SSBO
|
||||
{
|
||||
vec4 data[];
|
||||
} outdata;
|
||||
|
||||
layout(binding = 2, std430) readonly buffer SSBO3
|
||||
{
|
||||
Foo foos[];
|
||||
} foobar;
|
||||
|
||||
void baz(inout Foo foo)
|
||||
{
|
||||
uint ident = gl_GlobalInvocationID.x;
|
||||
foo.a = indata.data[(4u * ident) + 0u];
|
||||
foo.b = indata.data[(4u * ident) + 1u];
|
||||
foo.c = indata.data[(4u * ident) + 2u];
|
||||
foo.d = indata.data[(4u * ident) + 3u];
|
||||
}
|
||||
|
||||
void meow(inout Foo foo)
|
||||
{
|
||||
foo.a += vec4(10.0);
|
||||
foo.b += vec4(20.0);
|
||||
foo.c += vec4(30.0);
|
||||
foo.d += vec4(40.0);
|
||||
}
|
||||
|
||||
vec4 bar(Foo foo)
|
||||
{
|
||||
return ((foo.a + foo.b) + foo.c) + foo.d;
|
||||
}
|
||||
|
||||
void main()
|
||||
{
|
||||
Foo param;
|
||||
baz(param);
|
||||
Foo foo = param;
|
||||
Foo param_1 = foo;
|
||||
meow(param_1);
|
||||
foo = param_1;
|
||||
Foo param_2 = foo;
|
||||
Foo param_3;
|
||||
param_3.a = foobar.foos[gl_GlobalInvocationID.x].a;
|
||||
param_3.b = foobar.foos[gl_GlobalInvocationID.x].b;
|
||||
param_3.c = foobar.foos[gl_GlobalInvocationID.x].c;
|
||||
param_3.d = foobar.foos[gl_GlobalInvocationID.x].d;
|
||||
outdata.data[gl_GlobalInvocationID.x] = bar(param_2) + bar(param_3);
|
||||
}
|
||||
|
||||
|
|
@ -1,159 +0,0 @@
|
|||
#version 450
|
||||
#if defined(GL_AMD_gpu_shader_half_float)
|
||||
#extension GL_AMD_gpu_shader_half_float : require
|
||||
#elif defined(GL_NV_gpu_shader5)
|
||||
#extension GL_NV_gpu_shader5 : require
|
||||
#else
|
||||
#error No extension available for FP16.
|
||||
#endif
|
||||
|
||||
struct ResType
|
||||
{
|
||||
f16vec4 _m0;
|
||||
ivec4 _m1;
|
||||
};
|
||||
|
||||
layout(location = 3) in f16vec4 v4;
|
||||
layout(location = 2) in f16vec3 v3;
|
||||
layout(location = 0) in float16_t v1;
|
||||
layout(location = 1) in f16vec2 v2;
|
||||
|
||||
f16mat2 test_mat2(f16vec2 a, f16vec2 b, f16vec2 c, f16vec2 d)
|
||||
{
|
||||
return f16mat2(f16vec2(a), f16vec2(b)) * f16mat2(f16vec2(c), f16vec2(d));
|
||||
}
|
||||
|
||||
f16mat3 test_mat3(f16vec3 a, f16vec3 b, f16vec3 c, f16vec3 d, f16vec3 e, f16vec3 f)
|
||||
{
|
||||
return f16mat3(f16vec3(a), f16vec3(b), f16vec3(c)) * f16mat3(f16vec3(d), f16vec3(e), f16vec3(f));
|
||||
}
|
||||
|
||||
void test_constants()
|
||||
{
|
||||
float16_t a = float16_t(1.0);
|
||||
float16_t b = float16_t(1.5);
|
||||
float16_t c = float16_t(-1.5);
|
||||
float16_t d = float16_t(0.0 / 0.0);
|
||||
float16_t e = float16_t(1.0 / 0.0);
|
||||
float16_t f = float16_t(-1.0 / 0.0);
|
||||
float16_t g = float16_t(1014.0);
|
||||
float16_t h = float16_t(9.5367431640625e-07);
|
||||
}
|
||||
|
||||
float16_t test_result()
|
||||
{
|
||||
return float16_t(1.0);
|
||||
}
|
||||
|
||||
void test_conversions()
|
||||
{
|
||||
float16_t one = test_result();
|
||||
int a = int(one);
|
||||
uint b = uint(one);
|
||||
bool c = one != float16_t(0.0);
|
||||
float d = float(one);
|
||||
double e = double(one);
|
||||
float16_t a2 = float16_t(a);
|
||||
float16_t b2 = float16_t(b);
|
||||
float16_t c2 = float16_t(c);
|
||||
float16_t d2 = float16_t(d);
|
||||
float16_t e2 = float16_t(e);
|
||||
}
|
||||
|
||||
void test_builtins()
|
||||
{
|
||||
f16vec4 res = radians(v4);
|
||||
res = degrees(v4);
|
||||
res = sin(v4);
|
||||
res = cos(v4);
|
||||
res = tan(v4);
|
||||
res = asin(v4);
|
||||
res = atan(v4, v3.xyzz);
|
||||
res = atan(v4);
|
||||
res = sinh(v4);
|
||||
res = cosh(v4);
|
||||
res = tanh(v4);
|
||||
res = asinh(v4);
|
||||
res = acosh(v4);
|
||||
res = atanh(v4);
|
||||
res = pow(v4, v4);
|
||||
res = exp(v4);
|
||||
res = log(v4);
|
||||
res = exp2(v4);
|
||||
res = log2(v4);
|
||||
res = sqrt(v4);
|
||||
res = inversesqrt(v4);
|
||||
res = abs(v4);
|
||||
res = sign(v4);
|
||||
res = floor(v4);
|
||||
res = trunc(v4);
|
||||
res = round(v4);
|
||||
res = roundEven(v4);
|
||||
res = ceil(v4);
|
||||
res = fract(v4);
|
||||
res = mod(v4, v4);
|
||||
f16vec4 tmp;
|
||||
f16vec4 _231 = modf(v4, tmp);
|
||||
res = _231;
|
||||
res = min(v4, v4);
|
||||
res = max(v4, v4);
|
||||
res = clamp(v4, v4, v4);
|
||||
res = mix(v4, v4, v4);
|
||||
res = mix(v4, v4, lessThan(v4, v4));
|
||||
res = step(v4, v4);
|
||||
res = smoothstep(v4, v4, v4);
|
||||
bvec4 btmp = isnan(v4);
|
||||
btmp = isinf(v4);
|
||||
res = fma(v4, v4, v4);
|
||||
ResType _275;
|
||||
_275._m0 = frexp(v4, _275._m1);
|
||||
ivec4 itmp = _275._m1;
|
||||
res = _275._m0;
|
||||
res = ldexp(res, itmp);
|
||||
uint pack0 = packFloat2x16(v4.xy);
|
||||
uint pack1 = packFloat2x16(v4.zw);
|
||||
res = f16vec4(unpackFloat2x16(pack0), unpackFloat2x16(pack1));
|
||||
float16_t t0 = length(v4);
|
||||
t0 = distance(v4, v4);
|
||||
t0 = dot(v4, v4);
|
||||
f16vec3 res3 = cross(v3, v3);
|
||||
res = normalize(v4);
|
||||
res = faceforward(v4, v4, v4);
|
||||
res = reflect(v4, v4);
|
||||
res = refract(v4, v4, v1);
|
||||
btmp = lessThan(v4, v4);
|
||||
btmp = lessThanEqual(v4, v4);
|
||||
btmp = greaterThan(v4, v4);
|
||||
btmp = greaterThanEqual(v4, v4);
|
||||
btmp = equal(v4, v4);
|
||||
btmp = notEqual(v4, v4);
|
||||
res = dFdx(v4);
|
||||
res = dFdy(v4);
|
||||
res = dFdxFine(v4);
|
||||
res = dFdyFine(v4);
|
||||
res = dFdxCoarse(v4);
|
||||
res = dFdyCoarse(v4);
|
||||
res = fwidth(v4);
|
||||
res = fwidthFine(v4);
|
||||
res = fwidthCoarse(v4);
|
||||
}
|
||||
|
||||
void main()
|
||||
{
|
||||
f16vec2 param = v2;
|
||||
f16vec2 param_1 = v2;
|
||||
f16vec2 param_2 = v3.xy;
|
||||
f16vec2 param_3 = v3.xy;
|
||||
f16mat2 m0 = test_mat2(param, param_1, param_2, param_3);
|
||||
f16vec3 param_4 = v3;
|
||||
f16vec3 param_5 = v3;
|
||||
f16vec3 param_6 = v3;
|
||||
f16vec3 param_7 = v4.xyz;
|
||||
f16vec3 param_8 = v4.xyz;
|
||||
f16vec3 param_9 = v4.yzw;
|
||||
f16mat3 m1 = test_mat3(param_4, param_5, param_6, param_7, param_8, param_9);
|
||||
test_constants();
|
||||
test_conversions();
|
||||
test_builtins();
|
||||
}
|
||||
|
||||
|
|
@ -1,24 +0,0 @@
|
|||
#version 450
|
||||
|
||||
layout(binding = 0) uniform sampler2D uTextures[2 * 3 * 1];
|
||||
|
||||
layout(location = 1) in vec2 vUV;
|
||||
layout(location = 0) out vec4 FragColor;
|
||||
layout(location = 0) flat in int vIndex;
|
||||
|
||||
void main()
|
||||
{
|
||||
vec4 values3[2 * 3 * 1];
|
||||
for (int z = 0; z < 2; z++)
|
||||
{
|
||||
for (int y = 0; y < 3; y++)
|
||||
{
|
||||
for (int x = 0; x < 1; x++)
|
||||
{
|
||||
values3[z * 3 * 1 + y * 1 + x] = texture(uTextures[z * 3 * 1 + y * 1 + x], vUV);
|
||||
}
|
||||
}
|
||||
}
|
||||
FragColor = (values3[1 * 3 * 1 + 2 * 1 + 0] + values3[0 * 3 * 1 + 2 * 1 + 0]) + values3[(vIndex + 1) * 3 * 1 + 2 * 1 + vIndex];
|
||||
}
|
||||
|
||||
|
|
@ -1,25 +0,0 @@
|
|||
#version 450
|
||||
#if defined(GL_AMD_gpu_shader_half_float)
|
||||
#extension GL_AMD_gpu_shader_half_float : require
|
||||
#elif defined(GL_NV_gpu_shader5)
|
||||
#extension GL_NV_gpu_shader5 : require
|
||||
#else
|
||||
#error No extension available for FP16.
|
||||
#endif
|
||||
#if defined(GL_AMD_gpu_shader_int16)
|
||||
#extension GL_AMD_gpu_shader_int16 : require
|
||||
#else
|
||||
#error No extension available for Int16.
|
||||
#endif
|
||||
|
||||
layout(location = 0) out float16_t foo;
|
||||
layout(location = 1) out int16_t bar;
|
||||
layout(location = 2) out uint16_t baz;
|
||||
|
||||
void main()
|
||||
{
|
||||
foo = float16_t(1.0);
|
||||
bar = 2s;
|
||||
baz = 3us;
|
||||
}
|
||||
|
||||
|
|
@ -1,147 +0,0 @@
|
|||
#version 310 es
|
||||
#extension GL_EXT_scalar_block_layout : require
|
||||
layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
|
||||
|
||||
struct S0
|
||||
{
|
||||
vec2 a[1];
|
||||
float b;
|
||||
};
|
||||
|
||||
struct S1
|
||||
{
|
||||
vec3 a;
|
||||
float b;
|
||||
};
|
||||
|
||||
struct S2
|
||||
{
|
||||
vec3 a[1];
|
||||
float b;
|
||||
};
|
||||
|
||||
struct S3
|
||||
{
|
||||
vec2 a;
|
||||
float b;
|
||||
};
|
||||
|
||||
struct S4
|
||||
{
|
||||
vec2 c;
|
||||
};
|
||||
|
||||
struct Content
|
||||
{
|
||||
S0 m0s[1];
|
||||
S1 m1s[1];
|
||||
S2 m2s[1];
|
||||
S0 m0;
|
||||
S1 m1;
|
||||
S2 m2;
|
||||
S3 m3;
|
||||
float m4;
|
||||
S4 m3s[8];
|
||||
};
|
||||
|
||||
struct S0_1
|
||||
{
|
||||
vec2 a[1];
|
||||
float b;
|
||||
};
|
||||
|
||||
struct S1_1
|
||||
{
|
||||
vec3 a;
|
||||
float b;
|
||||
};
|
||||
|
||||
struct S2_1
|
||||
{
|
||||
vec3 a[1];
|
||||
float b;
|
||||
};
|
||||
|
||||
struct S3_1
|
||||
{
|
||||
vec2 a;
|
||||
float b;
|
||||
};
|
||||
|
||||
struct S4_1
|
||||
{
|
||||
vec2 c;
|
||||
};
|
||||
|
||||
struct Content_1
|
||||
{
|
||||
S0_1 m0s[1];
|
||||
S1_1 m1s[1];
|
||||
S2_1 m2s[1];
|
||||
S0_1 m0;
|
||||
S1_1 m1;
|
||||
S2_1 m2;
|
||||
S3_1 m3;
|
||||
float m4;
|
||||
S4_1 m3s[8];
|
||||
};
|
||||
|
||||
layout(set = 0, binding = 1, scalar) restrict buffer SSBO1
|
||||
{
|
||||
Content content;
|
||||
Content content1[2];
|
||||
Content content2;
|
||||
mat2 m0;
|
||||
mat2 m1;
|
||||
mat2x3 m2[4];
|
||||
mat3x2 m3;
|
||||
layout(row_major) mat2 m4;
|
||||
layout(row_major) mat2 m5[9];
|
||||
layout(row_major) mat2x3 m6[4][2];
|
||||
layout(row_major) mat3x2 m7;
|
||||
float array[];
|
||||
} ssbo_430;
|
||||
|
||||
layout(set = 0, binding = 0, std140) restrict buffer SSBO0
|
||||
{
|
||||
Content_1 content;
|
||||
Content_1 content1[2];
|
||||
Content_1 content2;
|
||||
mat2 m0;
|
||||
mat2 m1;
|
||||
mat2x3 m2[4];
|
||||
mat3x2 m3;
|
||||
layout(row_major) mat2 m4;
|
||||
layout(row_major) mat2 m5[9];
|
||||
layout(row_major) mat2x3 m6[4][2];
|
||||
layout(row_major) mat3x2 m7;
|
||||
float array[];
|
||||
} ssbo_140;
|
||||
|
||||
void main()
|
||||
{
|
||||
ssbo_430.content.m0s[0].a[0] = ssbo_140.content.m0s[0].a[0];
|
||||
ssbo_430.content.m0s[0].b = ssbo_140.content.m0s[0].b;
|
||||
ssbo_430.content.m1s[0].a = ssbo_140.content.m1s[0].a;
|
||||
ssbo_430.content.m1s[0].b = ssbo_140.content.m1s[0].b;
|
||||
ssbo_430.content.m2s[0].a[0] = ssbo_140.content.m2s[0].a[0];
|
||||
ssbo_430.content.m2s[0].b = ssbo_140.content.m2s[0].b;
|
||||
ssbo_430.content.m0.a[0] = ssbo_140.content.m0.a[0];
|
||||
ssbo_430.content.m0.b = ssbo_140.content.m0.b;
|
||||
ssbo_430.content.m1.a = ssbo_140.content.m1.a;
|
||||
ssbo_430.content.m1.b = ssbo_140.content.m1.b;
|
||||
ssbo_430.content.m2.a[0] = ssbo_140.content.m2.a[0];
|
||||
ssbo_430.content.m2.b = ssbo_140.content.m2.b;
|
||||
ssbo_430.content.m3.a = ssbo_140.content.m3.a;
|
||||
ssbo_430.content.m3.b = ssbo_140.content.m3.b;
|
||||
ssbo_430.content.m4 = ssbo_140.content.m4;
|
||||
ssbo_430.content.m3s[0].c = ssbo_140.content.m3s[0].c;
|
||||
ssbo_430.content.m3s[1].c = ssbo_140.content.m3s[1].c;
|
||||
ssbo_430.content.m3s[2].c = ssbo_140.content.m3s[2].c;
|
||||
ssbo_430.content.m3s[3].c = ssbo_140.content.m3s[3].c;
|
||||
ssbo_430.content.m3s[4].c = ssbo_140.content.m3s[4].c;
|
||||
ssbo_430.content.m3s[5].c = ssbo_140.content.m3s[5].c;
|
||||
ssbo_430.content.m3s[6].c = ssbo_140.content.m3s[6].c;
|
||||
ssbo_430.content.m3s[7].c = ssbo_140.content.m3s[7].c;
|
||||
}
|
||||
|
||||
|
|
@ -1,110 +0,0 @@
|
|||
#version 450
|
||||
#extension GL_KHR_shader_subgroup_basic : require
|
||||
#extension GL_KHR_shader_subgroup_ballot : require
|
||||
#extension GL_KHR_shader_subgroup_shuffle : require
|
||||
#extension GL_KHR_shader_subgroup_shuffle_relative : require
|
||||
#extension GL_KHR_shader_subgroup_vote : require
|
||||
#extension GL_KHR_shader_subgroup_arithmetic : require
|
||||
#extension GL_KHR_shader_subgroup_clustered : require
|
||||
#extension GL_KHR_shader_subgroup_quad : require
|
||||
layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
|
||||
|
||||
layout(set = 0, binding = 0, std430) buffer SSBO
|
||||
{
|
||||
float FragColor;
|
||||
} _9;
|
||||
|
||||
void main()
|
||||
{
|
||||
_9.FragColor = float(gl_NumSubgroups);
|
||||
_9.FragColor = float(gl_SubgroupID);
|
||||
_9.FragColor = float(gl_SubgroupSize);
|
||||
_9.FragColor = float(gl_SubgroupInvocationID);
|
||||
subgroupMemoryBarrier();
|
||||
subgroupBarrier();
|
||||
subgroupMemoryBarrier();
|
||||
subgroupMemoryBarrierBuffer();
|
||||
subgroupMemoryBarrierShared();
|
||||
subgroupMemoryBarrierImage();
|
||||
bool elected = subgroupElect();
|
||||
_9.FragColor = vec4(gl_SubgroupEqMask).x;
|
||||
_9.FragColor = vec4(gl_SubgroupGeMask).x;
|
||||
_9.FragColor = vec4(gl_SubgroupGtMask).x;
|
||||
_9.FragColor = vec4(gl_SubgroupLeMask).x;
|
||||
_9.FragColor = vec4(gl_SubgroupLtMask).x;
|
||||
vec4 broadcasted = subgroupBroadcast(vec4(10.0), 8u);
|
||||
vec3 first = subgroupBroadcastFirst(vec3(20.0));
|
||||
uvec4 ballot_value = subgroupBallot(true);
|
||||
bool inverse_ballot_value = subgroupInverseBallot(ballot_value);
|
||||
bool bit_extracted = subgroupBallotBitExtract(uvec4(10u), 8u);
|
||||
uint bit_count = subgroupBallotBitCount(ballot_value);
|
||||
uint inclusive_bit_count = subgroupBallotInclusiveBitCount(ballot_value);
|
||||
uint exclusive_bit_count = subgroupBallotExclusiveBitCount(ballot_value);
|
||||
uint lsb = subgroupBallotFindLSB(ballot_value);
|
||||
uint msb = subgroupBallotFindMSB(ballot_value);
|
||||
uint shuffled = subgroupShuffle(10u, 8u);
|
||||
uint shuffled_xor = subgroupShuffleXor(30u, 8u);
|
||||
uint shuffled_up = subgroupShuffleUp(20u, 4u);
|
||||
uint shuffled_down = subgroupShuffleDown(20u, 4u);
|
||||
bool has_all = subgroupAll(true);
|
||||
bool has_any = subgroupAny(true);
|
||||
bool has_equal = subgroupAllEqual(true);
|
||||
vec4 added = subgroupAdd(vec4(20.0));
|
||||
ivec4 iadded = subgroupAdd(ivec4(20));
|
||||
vec4 multiplied = subgroupMul(vec4(20.0));
|
||||
ivec4 imultiplied = subgroupMul(ivec4(20));
|
||||
vec4 lo = subgroupMin(vec4(20.0));
|
||||
vec4 hi = subgroupMax(vec4(20.0));
|
||||
ivec4 slo = subgroupMin(ivec4(20));
|
||||
ivec4 shi = subgroupMax(ivec4(20));
|
||||
uvec4 ulo = subgroupMin(uvec4(20u));
|
||||
uvec4 uhi = subgroupMax(uvec4(20u));
|
||||
uvec4 anded = subgroupAnd(ballot_value);
|
||||
uvec4 ored = subgroupOr(ballot_value);
|
||||
uvec4 xored = subgroupXor(ballot_value);
|
||||
added = subgroupInclusiveAdd(added);
|
||||
iadded = subgroupInclusiveAdd(iadded);
|
||||
multiplied = subgroupInclusiveMul(multiplied);
|
||||
imultiplied = subgroupInclusiveMul(imultiplied);
|
||||
lo = subgroupInclusiveMin(lo);
|
||||
hi = subgroupInclusiveMax(hi);
|
||||
slo = subgroupInclusiveMin(slo);
|
||||
shi = subgroupInclusiveMax(shi);
|
||||
ulo = subgroupInclusiveMin(ulo);
|
||||
uhi = subgroupInclusiveMax(uhi);
|
||||
anded = subgroupInclusiveAnd(anded);
|
||||
ored = subgroupInclusiveOr(ored);
|
||||
xored = subgroupInclusiveXor(ored);
|
||||
added = subgroupExclusiveAdd(lo);
|
||||
added = subgroupExclusiveAdd(multiplied);
|
||||
multiplied = subgroupExclusiveMul(multiplied);
|
||||
iadded = subgroupExclusiveAdd(imultiplied);
|
||||
imultiplied = subgroupExclusiveMul(imultiplied);
|
||||
lo = subgroupExclusiveMin(lo);
|
||||
hi = subgroupExclusiveMax(hi);
|
||||
ulo = subgroupExclusiveMin(ulo);
|
||||
uhi = subgroupExclusiveMax(uhi);
|
||||
slo = subgroupExclusiveMin(slo);
|
||||
shi = subgroupExclusiveMax(shi);
|
||||
anded = subgroupExclusiveAnd(anded);
|
||||
ored = subgroupExclusiveOr(ored);
|
||||
xored = subgroupExclusiveXor(ored);
|
||||
added = subgroupClusteredAdd(added, 4u);
|
||||
multiplied = subgroupClusteredMul(multiplied, 4u);
|
||||
iadded = subgroupClusteredAdd(iadded, 4u);
|
||||
imultiplied = subgroupClusteredMul(imultiplied, 4u);
|
||||
lo = subgroupClusteredMin(lo, 4u);
|
||||
hi = subgroupClusteredMax(hi, 4u);
|
||||
ulo = subgroupClusteredMin(ulo, 4u);
|
||||
uhi = subgroupClusteredMax(uhi, 4u);
|
||||
slo = subgroupClusteredMin(slo, 4u);
|
||||
shi = subgroupClusteredMax(shi, 4u);
|
||||
anded = subgroupClusteredAnd(anded, 4u);
|
||||
ored = subgroupClusteredOr(ored, 4u);
|
||||
xored = subgroupClusteredXor(xored, 4u);
|
||||
vec4 swap_horiz = subgroupQuadSwapHorizontal(vec4(20.0));
|
||||
vec4 swap_vertical = subgroupQuadSwapVertical(vec4(20.0));
|
||||
vec4 swap_diagonal = subgroupQuadSwapDiagonal(vec4(20.0));
|
||||
vec4 quad_broadcast = subgroupQuadBroadcast(vec4(20.0), 3u);
|
||||
}
|
||||
|
||||
|
|
@ -1,24 +0,0 @@
|
|||
#version 450
|
||||
#extension GL_EXT_scalar_block_layout : require
|
||||
|
||||
layout(set = 0, binding = 0, std430) uniform UBO
|
||||
{
|
||||
float a[1024];
|
||||
vec3 b[2];
|
||||
} _17;
|
||||
|
||||
layout(set = 0, binding = 1, std430) uniform UBOEnhancedLayout
|
||||
{
|
||||
layout(offset = 0) float c[1024];
|
||||
layout(offset = 4096) vec3 d[2];
|
||||
layout(offset = 10000) float e;
|
||||
} _30;
|
||||
|
||||
layout(location = 0) out float FragColor;
|
||||
layout(location = 0) flat in int vIndex;
|
||||
|
||||
void main()
|
||||
{
|
||||
FragColor = (_17.a[vIndex] + _30.c[vIndex]) + _30.e;
|
||||
}
|
||||
|
||||
Some files were not shown because too many files have changed in this diff Show More
Loading…
Reference in New Issue