diff --git a/3rdparty/spirv-cross/.clang-format b/3rdparty/spirv-cross/.clang-format new file mode 100755 index 000000000..443f90b77 --- /dev/null +++ b/3rdparty/spirv-cross/.clang-format @@ -0,0 +1,167 @@ +# The style used for all options not specifically set in the configuration. +BasedOnStyle: LLVM + +# The extra indent or outdent of access modifiers, e.g. public:. +AccessModifierOffset: -4 + +# If true, aligns escaped newlines as far left as possible. Otherwise puts them into the right-most column. +AlignEscapedNewlinesLeft: true + +# If true, aligns trailing comments. +AlignTrailingComments: false + +# Allow putting all parameters of a function declaration onto the next line even if BinPackParameters is false. +AllowAllParametersOfDeclarationOnNextLine: false + +# Allows contracting simple braced statements to a single line. +AllowShortBlocksOnASingleLine: false + +# If true, short case labels will be contracted to a single line. +AllowShortCaseLabelsOnASingleLine: false + +# Dependent on the value, int f() { return 0; } can be put on a single line. Possible values: None, Inline, All. +AllowShortFunctionsOnASingleLine: None + +# If true, if (a) return; can be put on a single line. +AllowShortIfStatementsOnASingleLine: false + +# If true, while (true) continue; can be put on a single line. +AllowShortLoopsOnASingleLine: false + +# If true, always break after function definition return types. +AlwaysBreakAfterDefinitionReturnType: false + +# If true, always break before multiline string literals. +AlwaysBreakBeforeMultilineStrings: false + +# If true, always break after the template<...> of a template declaration. +AlwaysBreakTemplateDeclarations: true + +# If false, a function call's arguments will either be all on the same line or will have one line each. +BinPackArguments: true + +# If false, a function declaration's or function definition's parameters will either all be on the same line +# or will have one line each. +BinPackParameters: true + +# The way to wrap binary operators. Possible values: None, NonAssignment, All. +BreakBeforeBinaryOperators: None + +# The brace breaking style to use. Possible values: Attach, Linux, Stroustrup, Allman, GNU. +BreakBeforeBraces: Allman + +# If true, ternary operators will be placed after line breaks. +BreakBeforeTernaryOperators: false + +# Always break constructor initializers before commas and align the commas with the colon. +BreakConstructorInitializersBeforeComma: true + +# The column limit. A column limit of 0 means that there is no column limit. +ColumnLimit: 120 + +# A regular expression that describes comments with special meaning, which should not be split into lines or otherwise changed. +CommentPragmas: '^ *' + +# If the constructor initializers don't fit on a line, put each initializer on its own line. +ConstructorInitializerAllOnOneLineOrOnePerLine: false + +# The number of characters to use for indentation of constructor initializer lists. +ConstructorInitializerIndentWidth: 4 + +# Indent width for line continuations. +ContinuationIndentWidth: 4 + +# If true, format braced lists as best suited for C++11 braced lists. +Cpp11BracedListStyle: false + +# Disables formatting at all. +DisableFormat: false + +# A vector of macros that should be interpreted as foreach loops instead of as function calls. +#ForEachMacros: '' + +# Indent case labels one level from the switch statement. +# When false, use the same indentation level as for the switch statement. +# Switch statement body is always indented one level more than case labels. +IndentCaseLabels: false + +# The number of columns to use for indentation. +IndentWidth: 4 + +# Indent if a function definition or declaration is wrapped after the type. +IndentWrappedFunctionNames: false + +# If true, empty lines at the start of blocks are kept. +KeepEmptyLinesAtTheStartOfBlocks: true + +# Language, this format style is targeted at. Possible values: None, Cpp, Java, JavaScript, Proto. +Language: Cpp + +# The maximum number of consecutive empty lines to keep. +MaxEmptyLinesToKeep: 1 + +# The indentation used for namespaces. Possible values: None, Inner, All. +NamespaceIndentation: None + +# The penalty for breaking a function call after "call(". +PenaltyBreakBeforeFirstCallParameter: 19 + +# The penalty for each line break introduced inside a comment. +PenaltyBreakComment: 300 + +# The penalty for breaking before the first <<. +PenaltyBreakFirstLessLess: 120 + +# The penalty for each line break introduced inside a string literal. +PenaltyBreakString: 1000 + +# The penalty for each character outside of the column limit. +PenaltyExcessCharacter: 1000000 + +# Penalty for putting the return type of a function onto its own line. +PenaltyReturnTypeOnItsOwnLine: 1000000000 + +# Pointer and reference alignment style. Possible values: Left, Right, Middle. +PointerAlignment: Right + +# If true, a space may be inserted after C style casts. +SpaceAfterCStyleCast: false + +# If false, spaces will be removed before assignment operators. +SpaceBeforeAssignmentOperators: true + +# Defines in which cases to put a space before opening parentheses. Possible values: Never, ControlStatements, Always. +SpaceBeforeParens: ControlStatements + +# If true, spaces may be inserted into '()'. +SpaceInEmptyParentheses: false + +# The number of spaces before trailing line comments (// - comments). +SpacesBeforeTrailingComments: 1 + +# If true, spaces will be inserted after '<' and before '>' in template argument lists. +SpacesInAngles: false + +# If true, spaces may be inserted into C style casts. +SpacesInCStyleCastParentheses: false + +# If true, spaces are inserted inside container literals (e.g. ObjC and Javascript array and dict literals). +SpacesInContainerLiterals: false + +# If true, spaces will be inserted after '(' and before ')'. +SpacesInParentheses: false + +# If true, spaces will be inserted after '[' and befor']'. +SpacesInSquareBrackets: false + +# Format compatible with this standard, e.g. use A > instead of A> for LS_Cpp03. Possible values: Cpp03, Cpp11, Auto. +Standard: Cpp11 + +# The number of columns used for tab stops. +TabWidth: 4 + +# The way to use tab characters in the resulting file. Possible values: Never, ForIndentation, Always. +UseTab: ForIndentation + +# Do not reflow comments +ReflowComments: false diff --git a/3rdparty/spirv-cross/.gitignore b/3rdparty/spirv-cross/.gitignore new file mode 100644 index 000000000..abd718958 --- /dev/null +++ b/3rdparty/spirv-cross/.gitignore @@ -0,0 +1,20 @@ +*.o +*.d +*.txt +/test +/spirv-cross +/obj +/msvc/x64 +/msvc/Debug +/msvc/Release +*.suo +*.sdf +*.opensdf +*.shader +*.a +*.bc +/external +.vs/ +*.vcxproj.user + +!CMakeLists.txt diff --git a/3rdparty/spirv-cross/.travis.yml b/3rdparty/spirv-cross/.travis.yml new file mode 100644 index 000000000..fa14abfde --- /dev/null +++ b/3rdparty/spirv-cross/.travis.yml @@ -0,0 +1,53 @@ +language: + - cpp + - python + +python: 3.7 + +matrix: + include: + - os: linux + dist: trusty + compiler: gcc + env: + - GENERATOR="Unix Makefiles" + - os: linux + dist: trusty + compiler: clang + env: + - GENERATOR="Unix Makefiles" + - os: osx + compiler: clang + osx_image: xcode10 + env: + - GENERATOR="Unix Makefiles" + - os: windows + before_install: + - choco install python3 + - choco install python2 + - export PATH="/c/Python27:/c/Python27/Scripts:$PATH" + - export PATH="/c/Python37:/c/Python37/Scripts:$PATH" + env: + - GENERATOR="Visual Studio 15 2017" + - os: windows + before_install: + - choco install python3 + - choco install python2 + - export PATH="/c/Python27:/c/Python27/Scripts:$PATH" + - export PATH="/c/Python37:/c/Python37/Scripts:$PATH" + env: + - GENERATOR="Visual Studio 15 2017 Win64" + +before_script: + - ./checkout_glslang_spirv_tools.sh + +script: + - if [[ "$TRAVIS_OS_NAME" == "windows" ]]; then PYTHON3=$(which python); fi + - if [[ "$TRAVIS_OS_NAME" != "windows" ]]; then PYTHON3=$(which python3); fi + - ./build_glslang_spirv_tools.sh Release + - mkdir build + - cd build + - cmake .. -DSPIRV_CROSS_SHARED=ON -DCMAKE_INSTALL_PREFIX=output -DCMAKE_BUILD_TYPE=Release -G "${GENERATOR}" -DPYTHON_EXECUTABLE:FILEPATH="${PYTHON3}" -DSPIRV_CROSS_ENABLE_TESTS=ON + - cmake --build . --config Release + - cmake --build . --config Release --target install + - ctest --verbose -C Release diff --git a/3rdparty/spirv-cross/checkout_glslang_spirv_tools.sh b/3rdparty/spirv-cross/checkout_glslang_spirv_tools.sh index 2ed5af3c5..ac8641de1 100755 --- a/3rdparty/spirv-cross/checkout_glslang_spirv_tools.sh +++ b/3rdparty/spirv-cross/checkout_glslang_spirv_tools.sh @@ -1,8 +1,8 @@ #!/bin/bash -GLSLANG_REV=a51d3d9f223361165127ded3cd2e59d81e22d91f -SPIRV_TOOLS_REV=bdcb155163d453fa67a03f59b4d5e00bd7c0f209 -SPIRV_HEADERS_REV=03a081524afabdde274d885880c2fef213e46a59 +GLSLANG_REV=ef807f4bc543e061f25dbbee6cb64dd5053b2adc +SPIRV_TOOLS_REV=12e4a7b649e6fe28683de9fc352200c82948a1f0 +SPIRV_HEADERS_REV=111a25e4ae45e2b4d7c18415e1d6884712b958c4 if [ -d external/glslang ]; then echo "Updating glslang to revision $GLSLANG_REV." diff --git a/3rdparty/spirv-cross/reference/opt/shaders-msl/asm/comp/variable-pointers-2.asm.comp b/3rdparty/spirv-cross/reference/opt/shaders-msl/asm/comp/variable-pointers-2.asm.comp index 77caa900e..b2dfc01b1 100644 --- a/3rdparty/spirv-cross/reference/opt/shaders-msl/asm/comp/variable-pointers-2.asm.comp +++ b/3rdparty/spirv-cross/reference/opt/shaders-msl/asm/comp/variable-pointers-2.asm.comp @@ -18,9 +18,14 @@ struct bar kernel void main0(device foo& buf [[buffer(0)]], constant bar& cb [[buffer(1)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]], uint3 gl_LocalInvocationID [[thread_position_in_threadgroup]]) { bool _70 = cb.d != 0; - for (device int* _52 = &(_70 ? &buf : nullptr)->a[0u], * _55 = &buf.a[0u]; (*_52) != (*_55); ) + device foo* _71 = _70 ? &buf : nullptr; + device foo* _67 = _71; + device foo* _45 = _71; + thread uint3* _77 = _70 ? &gl_GlobalInvocationID : &gl_LocalInvocationID; + thread uint3* _73 = _77; + for (device int* _52 = &_71->a[0u], * _55 = &buf.a[0u]; (*_52) != (*_55); ) { - int _66 = ((*_52) + (*_55)) + int((*(_70 ? &gl_GlobalInvocationID : &gl_LocalInvocationID)).x); + int _66 = ((*_52) + (*_55)) + int((*_77).x); *_52 = _66; *_55 = _66; _52 = &_52[1u]; diff --git a/3rdparty/spirv-cross/reference/opt/shaders-msl/asm/comp/variable-pointers-store-forwarding.asm.comp b/3rdparty/spirv-cross/reference/opt/shaders-msl/asm/comp/variable-pointers-store-forwarding.asm.comp index 9b29d85bd..b4e03a292 100644 --- a/3rdparty/spirv-cross/reference/opt/shaders-msl/asm/comp/variable-pointers-store-forwarding.asm.comp +++ b/3rdparty/spirv-cross/reference/opt/shaders-msl/asm/comp/variable-pointers-store-forwarding.asm.comp @@ -15,6 +15,11 @@ struct bar kernel void main0(device foo& x [[buffer(0)]], device bar& y [[buffer(1)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]]) { - y.b = x.a + x.a; + device int* _46 = (gl_GlobalInvocationID.x != 0u) ? &x.a : &y.b; + device int* _40 = _46; + device int* _33 = _46; + int _37 = x.a; + *_46 = 0; + y.b = _37 + _37; } diff --git a/3rdparty/spirv-cross/reference/opt/shaders-msl/comp/atomic.comp b/3rdparty/spirv-cross/reference/opt/shaders-msl/comp/atomic.comp index f77922aca..0315f3b12 100644 --- a/3rdparty/spirv-cross/reference/opt/shaders-msl/comp/atomic.comp +++ b/3rdparty/spirv-cross/reference/opt/shaders-msl/comp/atomic.comp @@ -27,7 +27,7 @@ kernel void main0(device SSBO& ssbo [[buffer(2)]]) do { _32 = 10u; - } while (!atomic_compare_exchange_weak_explicit((volatile device atomic_uint*)&ssbo.u32, &_32, 2u, memory_order_relaxed, memory_order_relaxed)); + } while (!atomic_compare_exchange_weak_explicit((volatile device atomic_uint*)&ssbo.u32, &_32, 2u, memory_order_relaxed, memory_order_relaxed) && _32 == 10u); int _36 = atomic_fetch_add_explicit((volatile device atomic_int*)&ssbo.i32, 1, memory_order_relaxed); int _38 = atomic_fetch_or_explicit((volatile device atomic_int*)&ssbo.i32, 1, memory_order_relaxed); int _40 = atomic_fetch_xor_explicit((volatile device atomic_int*)&ssbo.i32, 1, memory_order_relaxed); @@ -39,7 +39,7 @@ kernel void main0(device SSBO& ssbo [[buffer(2)]]) do { _52 = 10; - } while (!atomic_compare_exchange_weak_explicit((volatile device atomic_int*)&ssbo.i32, &_52, 2, memory_order_relaxed, memory_order_relaxed)); + } while (!atomic_compare_exchange_weak_explicit((volatile device atomic_int*)&ssbo.i32, &_52, 2, memory_order_relaxed, memory_order_relaxed) && _52 == 10); shared_u32 = 10u; shared_i32 = 10; uint _57 = atomic_fetch_add_explicit((volatile threadgroup atomic_uint*)&shared_u32, 1u, memory_order_relaxed); @@ -53,7 +53,7 @@ kernel void main0(device SSBO& ssbo [[buffer(2)]]) do { _64 = 10u; - } while (!atomic_compare_exchange_weak_explicit((volatile threadgroup atomic_uint*)&shared_u32, &_64, 2u, memory_order_relaxed, memory_order_relaxed)); + } while (!atomic_compare_exchange_weak_explicit((volatile threadgroup atomic_uint*)&shared_u32, &_64, 2u, memory_order_relaxed, memory_order_relaxed) && _64 == 10u); int _65 = atomic_fetch_add_explicit((volatile threadgroup atomic_int*)&shared_i32, 1, memory_order_relaxed); int _66 = atomic_fetch_or_explicit((volatile threadgroup atomic_int*)&shared_i32, 1, memory_order_relaxed); int _67 = atomic_fetch_xor_explicit((volatile threadgroup atomic_int*)&shared_i32, 1, memory_order_relaxed); @@ -65,6 +65,6 @@ kernel void main0(device SSBO& ssbo [[buffer(2)]]) do { _72 = 10; - } while (!atomic_compare_exchange_weak_explicit((volatile threadgroup atomic_int*)&shared_i32, &_72, 2, memory_order_relaxed, memory_order_relaxed)); + } while (!atomic_compare_exchange_weak_explicit((volatile threadgroup atomic_int*)&shared_i32, &_72, 2, memory_order_relaxed, memory_order_relaxed) && _72 == 10); } diff --git a/3rdparty/spirv-cross/reference/opt/shaders/asm/frag/do-while-statement-fallback.asm.frag b/3rdparty/spirv-cross/reference/opt/shaders/asm/frag/do-while-statement-fallback.asm.frag new file mode 100644 index 000000000..35bec9882 --- /dev/null +++ b/3rdparty/spirv-cross/reference/opt/shaders/asm/frag/do-while-statement-fallback.asm.frag @@ -0,0 +1,9 @@ +#version 450 + +layout(location = 0) out float FragColor; + +void main() +{ + FragColor = 5.0; +} + diff --git a/3rdparty/spirv-cross/reference/shaders-hlsl-no-opt/asm/frag/image-fetch-uint-coord.asm.frag b/3rdparty/spirv-cross/reference/shaders-hlsl-no-opt/asm/frag/image-fetch-uint-coord.asm.frag new file mode 100644 index 000000000..c5a761345 --- /dev/null +++ b/3rdparty/spirv-cross/reference/shaders-hlsl-no-opt/asm/frag/image-fetch-uint-coord.asm.frag @@ -0,0 +1,28 @@ +Texture2D Tex : register(t0); + +static uint3 in_var_TEXCOORD0; +static float4 out_var_SV_Target0; + +struct SPIRV_Cross_Input +{ + nointerpolation uint3 in_var_TEXCOORD0 : TEXCOORD0; +}; + +struct SPIRV_Cross_Output +{ + float4 out_var_SV_Target0 : SV_Target0; +}; + +void frag_main() +{ + out_var_SV_Target0 = Tex.Load(int3(in_var_TEXCOORD0.xy, in_var_TEXCOORD0.z)); +} + +SPIRV_Cross_Output main(SPIRV_Cross_Input stage_input) +{ + in_var_TEXCOORD0 = stage_input.in_var_TEXCOORD0; + frag_main(); + SPIRV_Cross_Output stage_output; + stage_output.out_var_SV_Target0 = out_var_SV_Target0; + return stage_output; +} diff --git a/3rdparty/spirv-cross/reference/shaders-msl-no-opt/asm/frag/image-fetch-uint-coord.asm.frag b/3rdparty/spirv-cross/reference/shaders-msl-no-opt/asm/frag/image-fetch-uint-coord.asm.frag new file mode 100644 index 000000000..887200f78 --- /dev/null +++ b/3rdparty/spirv-cross/reference/shaders-msl-no-opt/asm/frag/image-fetch-uint-coord.asm.frag @@ -0,0 +1,22 @@ +#include +#include + +using namespace metal; + +struct main0_out +{ + float4 out_var_SV_Target0 [[color(0)]]; +}; + +struct main0_in +{ + uint3 in_var_TEXCOORD0 [[user(locn0)]]; +}; + +fragment main0_out main0(main0_in in [[stage_in]], texture2d Tex [[texture(0)]]) +{ + main0_out out = {}; + out.out_var_SV_Target0 = Tex.read(uint2(in.in_var_TEXCOORD0.xy), in.in_var_TEXCOORD0.z); + return out; +} + diff --git a/3rdparty/spirv-cross/reference/shaders-msl-no-opt/asm/frag/image-type-normal-comparison-usage.asm.frag b/3rdparty/spirv-cross/reference/shaders-msl-no-opt/asm/frag/image-type-normal-comparison-usage.asm.frag new file mode 100644 index 000000000..2e43ab0c2 --- /dev/null +++ b/3rdparty/spirv-cross/reference/shaders-msl-no-opt/asm/frag/image-type-normal-comparison-usage.asm.frag @@ -0,0 +1,31 @@ +#include +#include + +using namespace metal; + +struct main0_out +{ + float4 out_var_SV_Target0 [[color(0)]]; +}; + +struct main0_in +{ + float2 in_var_TEXCOORD0 [[user(locn0)]]; +}; + +fragment main0_out main0(main0_in in [[stage_in]], depth2d ShadowMap [[texture(0)]], sampler SampleNormal [[sampler(0)]], sampler SampleShadow [[sampler(1)]]) +{ + main0_out out = {}; + float _41; + if (in.in_var_TEXCOORD0.x > 0.5) + { + _41 = float(float4(ShadowMap.sample(SampleNormal, in.in_var_TEXCOORD0)).x <= 0.5); + } + else + { + _41 = ShadowMap.sample_compare(SampleShadow, in.in_var_TEXCOORD0, 0.5, level(0.0)); + } + out.out_var_SV_Target0 = float4(_41, _41, _41, 1.0); + return out; +} + diff --git a/3rdparty/spirv-cross/reference/shaders-msl/comp/atomic.comp b/3rdparty/spirv-cross/reference/shaders-msl/comp/atomic.comp index f77922aca..0315f3b12 100644 --- a/3rdparty/spirv-cross/reference/shaders-msl/comp/atomic.comp +++ b/3rdparty/spirv-cross/reference/shaders-msl/comp/atomic.comp @@ -27,7 +27,7 @@ kernel void main0(device SSBO& ssbo [[buffer(2)]]) do { _32 = 10u; - } while (!atomic_compare_exchange_weak_explicit((volatile device atomic_uint*)&ssbo.u32, &_32, 2u, memory_order_relaxed, memory_order_relaxed)); + } while (!atomic_compare_exchange_weak_explicit((volatile device atomic_uint*)&ssbo.u32, &_32, 2u, memory_order_relaxed, memory_order_relaxed) && _32 == 10u); int _36 = atomic_fetch_add_explicit((volatile device atomic_int*)&ssbo.i32, 1, memory_order_relaxed); int _38 = atomic_fetch_or_explicit((volatile device atomic_int*)&ssbo.i32, 1, memory_order_relaxed); int _40 = atomic_fetch_xor_explicit((volatile device atomic_int*)&ssbo.i32, 1, memory_order_relaxed); @@ -39,7 +39,7 @@ kernel void main0(device SSBO& ssbo [[buffer(2)]]) do { _52 = 10; - } while (!atomic_compare_exchange_weak_explicit((volatile device atomic_int*)&ssbo.i32, &_52, 2, memory_order_relaxed, memory_order_relaxed)); + } while (!atomic_compare_exchange_weak_explicit((volatile device atomic_int*)&ssbo.i32, &_52, 2, memory_order_relaxed, memory_order_relaxed) && _52 == 10); shared_u32 = 10u; shared_i32 = 10; uint _57 = atomic_fetch_add_explicit((volatile threadgroup atomic_uint*)&shared_u32, 1u, memory_order_relaxed); @@ -53,7 +53,7 @@ kernel void main0(device SSBO& ssbo [[buffer(2)]]) do { _64 = 10u; - } while (!atomic_compare_exchange_weak_explicit((volatile threadgroup atomic_uint*)&shared_u32, &_64, 2u, memory_order_relaxed, memory_order_relaxed)); + } while (!atomic_compare_exchange_weak_explicit((volatile threadgroup atomic_uint*)&shared_u32, &_64, 2u, memory_order_relaxed, memory_order_relaxed) && _64 == 10u); int _65 = atomic_fetch_add_explicit((volatile threadgroup atomic_int*)&shared_i32, 1, memory_order_relaxed); int _66 = atomic_fetch_or_explicit((volatile threadgroup atomic_int*)&shared_i32, 1, memory_order_relaxed); int _67 = atomic_fetch_xor_explicit((volatile threadgroup atomic_int*)&shared_i32, 1, memory_order_relaxed); @@ -65,6 +65,6 @@ kernel void main0(device SSBO& ssbo [[buffer(2)]]) do { _72 = 10; - } while (!atomic_compare_exchange_weak_explicit((volatile threadgroup atomic_int*)&shared_i32, &_72, 2, memory_order_relaxed, memory_order_relaxed)); + } while (!atomic_compare_exchange_weak_explicit((volatile threadgroup atomic_int*)&shared_i32, &_72, 2, memory_order_relaxed, memory_order_relaxed) && _72 == 10); } diff --git a/3rdparty/spirv-cross/reference/shaders-no-opt/asm/frag/image-fetch-uint-coord.asm.frag b/3rdparty/spirv-cross/reference/shaders-no-opt/asm/frag/image-fetch-uint-coord.asm.frag new file mode 100644 index 000000000..8b8d0c858 --- /dev/null +++ b/3rdparty/spirv-cross/reference/shaders-no-opt/asm/frag/image-fetch-uint-coord.asm.frag @@ -0,0 +1,12 @@ +#version 450 + +uniform sampler2D SPIRV_Cross_CombinedTexSPIRV_Cross_DummySampler; + +layout(location = 0) flat in uvec3 in_var_TEXCOORD0; +layout(location = 0) out vec4 out_var_SV_Target0; + +void main() +{ + out_var_SV_Target0 = texelFetch(SPIRV_Cross_CombinedTexSPIRV_Cross_DummySampler, ivec2(in_var_TEXCOORD0.xy), int(in_var_TEXCOORD0.z)); +} + diff --git a/3rdparty/spirv-cross/reference/shaders/asm/frag/do-while-statement-fallback.asm.frag b/3rdparty/spirv-cross/reference/shaders/asm/frag/do-while-statement-fallback.asm.frag new file mode 100644 index 000000000..0e6947a39 --- /dev/null +++ b/3rdparty/spirv-cross/reference/shaders/asm/frag/do-while-statement-fallback.asm.frag @@ -0,0 +1,58 @@ +#version 450 + +layout(location = 0) out float FragColor; + +void main() +{ + float foo = 1.0; + for (;;) + { + foo = 2.0; + if (false) + { + continue; + } + else + { + break; + } + } + for (;;) + { + foo = 3.0; + if (false) + { + continue; + } + else + { + break; + } + } + for (;;) + { + foo = 4.0; + if (false) + { + continue; + } + else + { + break; + } + } + for (;;) + { + foo = 5.0; + if (false) + { + continue; + } + else + { + break; + } + } + FragColor = foo; +} + diff --git a/3rdparty/spirv-cross/shaders-hlsl-no-opt/asm/frag/image-fetch-uint-coord.asm.frag b/3rdparty/spirv-cross/shaders-hlsl-no-opt/asm/frag/image-fetch-uint-coord.asm.frag new file mode 100644 index 000000000..ca8022d4b --- /dev/null +++ b/3rdparty/spirv-cross/shaders-hlsl-no-opt/asm/frag/image-fetch-uint-coord.asm.frag @@ -0,0 +1,44 @@ +; SPIR-V +; Version: 1.0 +; Generator: Google spiregg; 0 +; Bound: 29 +; Schema: 0 + OpCapability Shader + OpMemoryModel Logical GLSL450 + OpEntryPoint Fragment %main "main" %in_var_TEXCOORD0 %out_var_SV_Target0 + OpExecutionMode %main OriginUpperLeft + OpSource HLSL 600 + OpName %type_2d_image "type.2d.image" + OpName %Tex "Tex" + OpName %in_var_TEXCOORD0 "in.var.TEXCOORD0" + OpName %out_var_SV_Target0 "out.var.SV_Target0" + OpName %main "main" + OpDecorate %in_var_TEXCOORD0 Flat + OpDecorate %in_var_TEXCOORD0 Location 0 + OpDecorate %out_var_SV_Target0 Location 0 + OpDecorate %Tex DescriptorSet 0 + OpDecorate %Tex Binding 0 + %int = OpTypeInt 32 1 + %int_2 = OpConstant %int 2 + %float = OpTypeFloat 32 +%type_2d_image = OpTypeImage %float 2D 2 0 0 1 Unknown +%_ptr_UniformConstant_type_2d_image = OpTypePointer UniformConstant %type_2d_image + %uint = OpTypeInt 32 0 + %v3uint = OpTypeVector %uint 3 +%_ptr_Input_v3uint = OpTypePointer Input %v3uint + %v4float = OpTypeVector %float 4 +%_ptr_Output_v4float = OpTypePointer Output %v4float + %void = OpTypeVoid + %16 = OpTypeFunction %void + %Tex = OpVariable %_ptr_UniformConstant_type_2d_image UniformConstant +%in_var_TEXCOORD0 = OpVariable %_ptr_Input_v3uint Input +%out_var_SV_Target0 = OpVariable %_ptr_Output_v4float Output + %main = OpFunction %void None %16 + %19 = OpLabel + %20 = OpLoad %v3uint %in_var_TEXCOORD0 + %21 = OpCompositeExtract %uint %20 2 + %27 = OpLoad %type_2d_image %Tex + %28 = OpImageFetch %v4float %27 %20 Lod %21 + OpStore %out_var_SV_Target0 %28 + OpReturn + OpFunctionEnd diff --git a/3rdparty/spirv-cross/shaders-msl-no-opt/asm/frag/image-fetch-uint-coord.asm.frag b/3rdparty/spirv-cross/shaders-msl-no-opt/asm/frag/image-fetch-uint-coord.asm.frag new file mode 100644 index 000000000..ca8022d4b --- /dev/null +++ b/3rdparty/spirv-cross/shaders-msl-no-opt/asm/frag/image-fetch-uint-coord.asm.frag @@ -0,0 +1,44 @@ +; SPIR-V +; Version: 1.0 +; Generator: Google spiregg; 0 +; Bound: 29 +; Schema: 0 + OpCapability Shader + OpMemoryModel Logical GLSL450 + OpEntryPoint Fragment %main "main" %in_var_TEXCOORD0 %out_var_SV_Target0 + OpExecutionMode %main OriginUpperLeft + OpSource HLSL 600 + OpName %type_2d_image "type.2d.image" + OpName %Tex "Tex" + OpName %in_var_TEXCOORD0 "in.var.TEXCOORD0" + OpName %out_var_SV_Target0 "out.var.SV_Target0" + OpName %main "main" + OpDecorate %in_var_TEXCOORD0 Flat + OpDecorate %in_var_TEXCOORD0 Location 0 + OpDecorate %out_var_SV_Target0 Location 0 + OpDecorate %Tex DescriptorSet 0 + OpDecorate %Tex Binding 0 + %int = OpTypeInt 32 1 + %int_2 = OpConstant %int 2 + %float = OpTypeFloat 32 +%type_2d_image = OpTypeImage %float 2D 2 0 0 1 Unknown +%_ptr_UniformConstant_type_2d_image = OpTypePointer UniformConstant %type_2d_image + %uint = OpTypeInt 32 0 + %v3uint = OpTypeVector %uint 3 +%_ptr_Input_v3uint = OpTypePointer Input %v3uint + %v4float = OpTypeVector %float 4 +%_ptr_Output_v4float = OpTypePointer Output %v4float + %void = OpTypeVoid + %16 = OpTypeFunction %void + %Tex = OpVariable %_ptr_UniformConstant_type_2d_image UniformConstant +%in_var_TEXCOORD0 = OpVariable %_ptr_Input_v3uint Input +%out_var_SV_Target0 = OpVariable %_ptr_Output_v4float Output + %main = OpFunction %void None %16 + %19 = OpLabel + %20 = OpLoad %v3uint %in_var_TEXCOORD0 + %21 = OpCompositeExtract %uint %20 2 + %27 = OpLoad %type_2d_image %Tex + %28 = OpImageFetch %v4float %27 %20 Lod %21 + OpStore %out_var_SV_Target0 %28 + OpReturn + OpFunctionEnd diff --git a/3rdparty/spirv-cross/shaders-msl-no-opt/asm/frag/image-type-normal-comparison-usage.asm.frag b/3rdparty/spirv-cross/shaders-msl-no-opt/asm/frag/image-type-normal-comparison-usage.asm.frag new file mode 100644 index 000000000..d4c264397 --- /dev/null +++ b/3rdparty/spirv-cross/shaders-msl-no-opt/asm/frag/image-type-normal-comparison-usage.asm.frag @@ -0,0 +1,76 @@ +; SPIR-V +; Version: 1.0 +; Generator: Google spiregg; 0 +; Bound: 43 +; Schema: 0 + OpCapability Shader + OpMemoryModel Logical GLSL450 + OpEntryPoint Fragment %main "main" %in_var_TEXCOORD0 %out_var_SV_Target0 + OpExecutionMode %main OriginUpperLeft + OpSource HLSL 600 + OpName %type_2d_image "type.2d.image" + OpName %ShadowMap "ShadowMap" + OpName %type_sampler "type.sampler" + OpName %SampleNormal "SampleNormal" + OpName %SampleShadow "SampleShadow" + OpName %in_var_TEXCOORD0 "in.var.TEXCOORD0" + OpName %out_var_SV_Target0 "out.var.SV_Target0" + OpName %main "main" + OpName %type_sampled_image "type.sampled.image" + OpDecorate %in_var_TEXCOORD0 Location 0 + OpDecorate %out_var_SV_Target0 Location 0 + OpDecorate %ShadowMap DescriptorSet 0 + OpDecorate %ShadowMap Binding 0 + OpDecorate %SampleNormal DescriptorSet 0 + OpDecorate %SampleNormal Binding 0 + OpDecorate %SampleShadow DescriptorSet 0 + OpDecorate %SampleShadow Binding 1 + %float = OpTypeFloat 32 + %float_0_5 = OpConstant %float 0.5 + %float_1 = OpConstant %float 1 + %float_0 = OpConstant %float 0 + %v4float = OpTypeVector %float 4 +%type_2d_image = OpTypeImage %float 2D 2 0 0 1 Unknown +%_ptr_UniformConstant_type_2d_image = OpTypePointer UniformConstant %type_2d_image +%type_sampler = OpTypeSampler +%_ptr_UniformConstant_type_sampler = OpTypePointer UniformConstant %type_sampler + %v2float = OpTypeVector %float 2 +%_ptr_Input_v2float = OpTypePointer Input %v2float +%_ptr_Output_v4float = OpTypePointer Output %v4float + %void = OpTypeVoid + %21 = OpTypeFunction %void + %bool = OpTypeBool +%type_sampled_image = OpTypeSampledImage %type_2d_image + %ShadowMap = OpVariable %_ptr_UniformConstant_type_2d_image UniformConstant +%SampleNormal = OpVariable %_ptr_UniformConstant_type_sampler UniformConstant +%SampleShadow = OpVariable %_ptr_UniformConstant_type_sampler UniformConstant +%in_var_TEXCOORD0 = OpVariable %_ptr_Input_v2float Input +%out_var_SV_Target0 = OpVariable %_ptr_Output_v4float Output + %main = OpFunction %void None %21 + %23 = OpLabel + %24 = OpLoad %v2float %in_var_TEXCOORD0 + %25 = OpCompositeExtract %float %24 0 + %26 = OpFOrdGreaterThan %bool %25 %float_0_5 + OpSelectionMerge %27 None + OpBranchConditional %26 %28 %29 + %28 = OpLabel + %30 = OpLoad %type_2d_image %ShadowMap + %31 = OpLoad %type_sampler %SampleNormal + %32 = OpSampledImage %type_sampled_image %30 %31 + %33 = OpImageSampleImplicitLod %v4float %32 %24 None + %34 = OpCompositeExtract %float %33 0 + %35 = OpFOrdLessThanEqual %bool %34 %float_0_5 + %36 = OpSelect %float %35 %float_1 %float_0 + OpBranch %27 + %29 = OpLabel + %37 = OpLoad %type_2d_image %ShadowMap + %38 = OpLoad %type_sampler %SampleShadow + %39 = OpSampledImage %type_sampled_image %37 %38 + %40 = OpImageSampleDrefExplicitLod %float %39 %24 %float_0_5 Lod %float_0 + OpBranch %27 + %27 = OpLabel + %41 = OpPhi %float %36 %28 %40 %29 + %42 = OpCompositeConstruct %v4float %41 %41 %41 %float_1 + OpStore %out_var_SV_Target0 %42 + OpReturn + OpFunctionEnd diff --git a/3rdparty/spirv-cross/shaders-no-opt/asm/frag/image-fetch-uint-coord.asm.frag b/3rdparty/spirv-cross/shaders-no-opt/asm/frag/image-fetch-uint-coord.asm.frag new file mode 100644 index 000000000..ca8022d4b --- /dev/null +++ b/3rdparty/spirv-cross/shaders-no-opt/asm/frag/image-fetch-uint-coord.asm.frag @@ -0,0 +1,44 @@ +; SPIR-V +; Version: 1.0 +; Generator: Google spiregg; 0 +; Bound: 29 +; Schema: 0 + OpCapability Shader + OpMemoryModel Logical GLSL450 + OpEntryPoint Fragment %main "main" %in_var_TEXCOORD0 %out_var_SV_Target0 + OpExecutionMode %main OriginUpperLeft + OpSource HLSL 600 + OpName %type_2d_image "type.2d.image" + OpName %Tex "Tex" + OpName %in_var_TEXCOORD0 "in.var.TEXCOORD0" + OpName %out_var_SV_Target0 "out.var.SV_Target0" + OpName %main "main" + OpDecorate %in_var_TEXCOORD0 Flat + OpDecorate %in_var_TEXCOORD0 Location 0 + OpDecorate %out_var_SV_Target0 Location 0 + OpDecorate %Tex DescriptorSet 0 + OpDecorate %Tex Binding 0 + %int = OpTypeInt 32 1 + %int_2 = OpConstant %int 2 + %float = OpTypeFloat 32 +%type_2d_image = OpTypeImage %float 2D 2 0 0 1 Unknown +%_ptr_UniformConstant_type_2d_image = OpTypePointer UniformConstant %type_2d_image + %uint = OpTypeInt 32 0 + %v3uint = OpTypeVector %uint 3 +%_ptr_Input_v3uint = OpTypePointer Input %v3uint + %v4float = OpTypeVector %float 4 +%_ptr_Output_v4float = OpTypePointer Output %v4float + %void = OpTypeVoid + %16 = OpTypeFunction %void + %Tex = OpVariable %_ptr_UniformConstant_type_2d_image UniformConstant +%in_var_TEXCOORD0 = OpVariable %_ptr_Input_v3uint Input +%out_var_SV_Target0 = OpVariable %_ptr_Output_v4float Output + %main = OpFunction %void None %16 + %19 = OpLabel + %20 = OpLoad %v3uint %in_var_TEXCOORD0 + %21 = OpCompositeExtract %uint %20 2 + %27 = OpLoad %type_2d_image %Tex + %28 = OpImageFetch %v4float %27 %20 Lod %21 + OpStore %out_var_SV_Target0 %28 + OpReturn + OpFunctionEnd diff --git a/3rdparty/spirv-cross/shaders/asm/frag/do-while-statement-fallback.asm.frag b/3rdparty/spirv-cross/shaders/asm/frag/do-while-statement-fallback.asm.frag new file mode 100644 index 000000000..08de3ef20 --- /dev/null +++ b/3rdparty/spirv-cross/shaders/asm/frag/do-while-statement-fallback.asm.frag @@ -0,0 +1,76 @@ +; SPIR-V +; Version: 1.0 +; Generator: Khronos Glslang Reference Front End; 7 +; Bound: 35 +; Schema: 0 + OpCapability Shader + %1 = OpExtInstImport "GLSL.std.450" + OpMemoryModel Logical GLSL450 + OpEntryPoint Fragment %main "main" %FragColor + OpExecutionMode %main OriginUpperLeft + OpSource GLSL 450 + OpName %main "main" + OpName %foo "foo" + OpName %FragColor "FragColor" + OpDecorate %FragColor Location 0 + %void = OpTypeVoid + %3 = OpTypeFunction %void + %float = OpTypeFloat 32 +%_ptr_Function_float = OpTypePointer Function %float + %float_1 = OpConstant %float 1 + %float_2 = OpConstant %float 2 + %bool = OpTypeBool + %false = OpConstantFalse %bool + %float_3 = OpConstant %float 3 + %float_4 = OpConstant %float 4 + %float_5 = OpConstant %float 5 +%_ptr_Output_float = OpTypePointer Output %float + %FragColor = OpVariable %_ptr_Output_float Output + %main = OpFunction %void None %3 + %5 = OpLabel + %foo = OpVariable %_ptr_Function_float Function + OpStore %foo %float_1 + OpBranch %10 + %10 = OpLabel + OpLoopMerge %12 %13 None + OpBranch %11 + %11 = OpLabel + OpBranch %13 + %13 = OpLabel + OpStore %foo %float_2 + OpBranchConditional %false %10 %12 + %12 = OpLabel + OpBranch %17 + %17 = OpLabel + OpLoopMerge %19 %20 None + OpBranch %18 + %18 = OpLabel + OpBranch %20 + %20 = OpLabel + OpStore %foo %float_3 + OpBranchConditional %false %17 %19 + %19 = OpLabel + OpBranch %22 + %22 = OpLabel + OpLoopMerge %24 %25 None + OpBranch %23 + %23 = OpLabel + OpBranch %25 + %25 = OpLabel + OpStore %foo %float_4 + OpBranchConditional %false %22 %24 + %24 = OpLabel + OpBranch %27 + %27 = OpLabel + OpLoopMerge %29 %30 None + OpBranch %28 + %28 = OpLabel + OpBranch %30 + %30 = OpLabel + OpStore %foo %float_5 + OpBranchConditional %false %27 %29 + %29 = OpLabel + %34 = OpLoad %float %foo + OpStore %FragColor %34 + OpReturn + OpFunctionEnd diff --git a/3rdparty/spirv-cross/spirv_cpp.cpp b/3rdparty/spirv-cross/spirv_cpp.cpp index 40424857a..a45a6b9ef 100644 --- a/3rdparty/spirv-cross/spirv_cpp.cpp +++ b/3rdparty/spirv-cross/spirv_cpp.cpp @@ -342,7 +342,7 @@ string CompilerCPP::compile() emit_function(get(ir.default_entry_point), Bitset()); pass_count++; - } while (force_recompile); + } while (is_forcing_recompilation()); // Match opening scope of emit_header(). end_scope_decl(); diff --git a/3rdparty/spirv-cross/spirv_cross.cpp b/3rdparty/spirv-cross/spirv_cross.cpp index f3a657f67..a913a1d4b 100644 --- a/3rdparty/spirv-cross/spirv_cross.cpp +++ b/3rdparty/spirv-cross/spirv_cross.cpp @@ -314,7 +314,7 @@ void Compiler::register_write(uint32_t chain) if (var->parameter && var->parameter->write_count == 0) { var->parameter->write_count++; - force_recompile = true; + force_recompile(); } } else @@ -1747,7 +1747,7 @@ uint32_t Compiler::type_struct_member_array_stride(const SPIRType &type, uint32_ SPIRV_CROSS_THROW("Struct member does not have ArrayStride set."); } else - SPIRV_CROSS_THROW("Struct member does not have Offset set."); + SPIRV_CROSS_THROW("Struct member does not have ArrayStride set."); } uint32_t Compiler::type_struct_member_matrix_stride(const SPIRType &type, uint32_t index) const @@ -4119,3 +4119,19 @@ bool Compiler::type_is_opaque_value(const SPIRType &type) const return !type.pointer && (type.basetype == SPIRType::SampledImage || type.basetype == SPIRType::Image || type.basetype == SPIRType::Sampler); } + +// Make these member functions so we can easily break on any force_recompile events. +void Compiler::force_recompile() +{ + is_force_recompile = true; +} + +bool Compiler::is_forcing_recompilation() const +{ + return is_force_recompile; +} + +void Compiler::clear_force_recompile() +{ + is_force_recompile = false; +} diff --git a/3rdparty/spirv-cross/spirv_cross.hpp b/3rdparty/spirv-cross/spirv_cross.hpp index 9ba43916f..fccea9d74 100644 --- a/3rdparty/spirv-cross/spirv_cross.hpp +++ b/3rdparty/spirv-cross/spirv_cross.hpp @@ -670,7 +670,10 @@ protected: bool execution_is_noop(const SPIRBlock &from, const SPIRBlock &to) const; SPIRBlock::ContinueBlockType continue_block_type(const SPIRBlock &continue_block) const; - bool force_recompile = false; + void force_recompile(); + void clear_force_recompile(); + bool is_forcing_recompilation() const; + bool is_force_recompile = false; bool block_is_loop_candidate(const SPIRBlock &block, SPIRBlock::Method method) const; diff --git a/3rdparty/spirv-cross/spirv_glsl.cpp b/3rdparty/spirv-cross/spirv_glsl.cpp index c8bf7e0e7..76e321348 100644 --- a/3rdparty/spirv-cross/spirv_glsl.cpp +++ b/3rdparty/spirv-cross/spirv_glsl.cpp @@ -288,7 +288,7 @@ void CompilerGLSL::reset() // We do some speculative optimizations which should pretty much always work out, // but just in case the SPIR-V is rather weird, recompile until it's happy. // This typically only means one extra pass. - force_recompile = false; + clear_force_recompile(); // Clear invalid expression tracking. invalid_expressions.clear(); @@ -463,7 +463,7 @@ string CompilerGLSL::compile() emit_function(get(ir.default_entry_point), Bitset()); pass_count++; - } while (force_recompile); + } while (is_forcing_recompilation()); // Entry point in GLSL is always main(). get_entry_point().name = "main"; @@ -2454,7 +2454,7 @@ void CompilerGLSL::handle_invalid_expression(uint32_t id) // We tried to read an invalidated expression. // This means we need another pass at compilation, but next time, force temporary variables so that they cannot be invalidated. forced_temporaries.insert(id); - force_recompile = true; + force_recompile(); } // Converts the format of the current expression from packed to unpacked, @@ -2665,7 +2665,7 @@ string CompilerGLSL::to_expression(uint32_t id, bool register_expression_read) } else { - if (force_recompile) + if (is_forcing_recompilation()) { // During first compilation phase, certain expression patterns can trigger exponential growth of memory. // Avoid this by returning dummy expressions during this phase. @@ -3598,7 +3598,7 @@ string CompilerGLSL::declare_temporary(uint32_t result_type, uint32_t result_id) { header.declare_temporary.emplace_back(result_type, result_id); hoisted_temporaries.insert(result_id); - force_recompile = true; + force_recompile(); } return join(to_name(result_id), " = "); @@ -4426,16 +4426,26 @@ void CompilerGLSL::emit_texture_op(const Instruction &i) // Sampling from a texture which was deduced to be a depth image, might actually return 1 component here. // Remap back to 4 components as sampling opcodes expect. - bool image_is_depth; - const auto *combined = maybe_get(img); - if (combined) - image_is_depth = image_is_comparison(imgtype, combined->image); - else - image_is_depth = image_is_comparison(imgtype, img); - - if (image_is_depth && backend.comparison_image_samples_scalar && image_opcode_is_sample_no_dref(op)) + if (backend.comparison_image_samples_scalar && image_opcode_is_sample_no_dref(op)) { - expr = remap_swizzle(get(result_type), 1, expr); + bool image_is_depth = false; + const auto *combined = maybe_get(img); + uint32_t image_id = combined ? combined->image : img; + + if (combined && image_is_comparison(imgtype, combined->image)) + image_is_depth = true; + else if (image_is_comparison(imgtype, img)) + image_is_depth = true; + + // We must also check the backing variable for the image. + // We might have loaded an OpImage, and used that handle for two different purposes. + // Once with comparison, once without. + auto *image_variable = maybe_get_backing_variable(image_id); + if (image_variable && image_is_comparison(get(image_variable->basetype), image_variable->self)) + image_is_depth = true; + + if (image_is_depth) + expr = remap_swizzle(get(result_type), 1, expr); } // Deals with reads from MSL. We might need to downconvert to fewer components. @@ -4590,6 +4600,7 @@ string CompilerGLSL::to_function_args(uint32_t img, const SPIRType &imgtype, boo if (coord_type.basetype == SPIRType::UInt) { auto expected_type = coord_type; + expected_type.vecsize = coord_components; expected_type.basetype = SPIRType::Int; coord_expr = bitcast_expression(expected_type, coord_type.basetype, coord_expr); } @@ -4690,7 +4701,19 @@ string CompilerGLSL::to_function_args(uint32_t img, const SPIRType &imgtype, boo { forward = forward && should_forward(lod); farg_str += ", "; - farg_str += to_expression(lod); + + auto &lod_expr_type = expression_type(lod); + + // Lod expression for TexelFetch in GLSL must be int, and only int. + if (is_fetch && imgtype.image.dim != DimBuffer && !imgtype.image.ms && + lod_expr_type.basetype != SPIRType::Int) + { + farg_str += join("int(", to_expression(lod), ")"); + } + else + { + farg_str += to_expression(lod); + } } } } @@ -6624,7 +6647,7 @@ void CompilerGLSL::track_expression_read(uint32_t id) forced_temporaries.insert(id); // Force a recompile after this pass to avoid forwarding this variable. - force_recompile = true; + force_recompile(); } } } @@ -6953,7 +6976,7 @@ void CompilerGLSL::disallow_forwarding_in_expression_chain(const SPIRExpression if (forwarded_temporaries.count(expr.self)) { forced_temporaries.insert(expr.self); - force_recompile = true; + force_recompile(); } for (auto &dependent : expr.expression_dependencies) @@ -8543,7 +8566,7 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction) if (flags.get(DecorationNonReadable)) { flags.clear(DecorationNonReadable); - force_recompile = true; + force_recompile(); } } @@ -8691,7 +8714,7 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction) if (flags.get(DecorationNonWritable)) { flags.clear(DecorationNonWritable); - force_recompile = true; + force_recompile(); } } @@ -9890,7 +9913,7 @@ void CompilerGLSL::require_extension_internal(const string &ext) if (backend.supports_extensions && !has_extension(ext)) { forced_extensions.push_back(ext); - force_recompile = true; + force_recompile(); } } @@ -9929,7 +9952,7 @@ bool CompilerGLSL::check_atomic_image(uint32_t id) { flags.clear(DecorationNonWritable); flags.clear(DecorationNonReadable); - force_recompile = true; + force_recompile(); } } return true; @@ -10237,7 +10260,7 @@ void CompilerGLSL::flush_phi(uint32_t from, uint32_t to) if (!var.allocate_temporary_copy) { var.allocate_temporary_copy = true; - force_recompile = true; + force_recompile(); } statement("_", phi.function_variable, "_copy", " = ", to_name(phi.function_variable), ";"); temporary_phi_variables.insert(phi.function_variable); @@ -10347,7 +10370,7 @@ void CompilerGLSL::branch(uint32_t from, uint32_t to) { if (!current_emitting_switch->need_ladder_break) { - force_recompile = true; + force_recompile(); current_emitting_switch->need_ladder_break = true; } @@ -10705,7 +10728,7 @@ bool CompilerGLSL::attempt_emit_loop_header(SPIRBlock &block, SPIRBlock::Method else { block.disable_block_optimization = true; - force_recompile = true; + force_recompile(); begin_scope(); // We'll see an end_scope() later. return false; } @@ -10781,7 +10804,7 @@ bool CompilerGLSL::attempt_emit_loop_header(SPIRBlock &block, SPIRBlock::Method else { block.disable_block_optimization = true; - force_recompile = true; + force_recompile(); begin_scope(); // We'll see an end_scope() later. return false; } @@ -10928,7 +10951,7 @@ void CompilerGLSL::emit_block_chain(SPIRBlock &block) // as writes to said loop variables might have been masked out, we need a recompile. if (!emitted_loop_header_variables && !block.loop_variables.empty()) { - force_recompile = true; + force_recompile(); for (auto var : block.loop_variables) get(var).loop_variable = false; block.loop_variables.clear(); @@ -11188,12 +11211,13 @@ void CompilerGLSL::emit_block_chain(SPIRBlock &block) bool positive_test = execution_is_noop(get(continue_block.true_block), get(continue_block.loop_dominator)); + uint32_t current_count = statement_count; auto statements = emit_continue_block(block.continue_block, positive_test, !positive_test); - if (!statements.empty()) + if (statement_count != current_count) { // The DoWhile block has side effects, force ComplexLoop pattern next pass. get(block.continue_block).complex_continue = true; - force_recompile = true; + force_recompile(); } // Might have to invert the do-while test here. diff --git a/3rdparty/spirv-cross/spirv_glsl.hpp b/3rdparty/spirv-cross/spirv_glsl.hpp index af61b3701..ca447cdeb 100644 --- a/3rdparty/spirv-cross/spirv_glsl.hpp +++ b/3rdparty/spirv-cross/spirv_glsl.hpp @@ -280,7 +280,7 @@ protected: template inline void statement(Ts &&... ts) { - if (force_recompile) + if (is_forcing_recompilation()) { // Do not bother emitting code while force_recompile is active. // We will compile again. @@ -289,7 +289,10 @@ protected: } if (redirect_statement) + { redirect_statement->push_back(join(std::forward(ts)...)); + statement_count++; + } else { for (uint32_t i = 0; i < indent; i++) diff --git a/3rdparty/spirv-cross/spirv_hlsl.cpp b/3rdparty/spirv-cross/spirv_hlsl.cpp index 0cbf75b27..c3b0f7905 100644 --- a/3rdparty/spirv-cross/spirv_hlsl.cpp +++ b/3rdparty/spirv-cross/spirv_hlsl.cpp @@ -2716,7 +2716,8 @@ void CompilerHLSL::emit_texture_op(const Instruction &i) // The IR can give us more components than we need, so chop them off as needed. string coord_expr; - if (coord_components != expression_type(coord).vecsize) + auto &coord_type = expression_type(coord); + if (coord_components != coord_type.vecsize) coord_expr = to_enclosed_expression(coord) + swizzle(coord_components, expression_type(coord).vecsize); else coord_expr = to_expression(coord); @@ -2726,9 +2727,8 @@ void CompilerHLSL::emit_texture_op(const Instruction &i) if (hlsl_options.shader_model < 40 && lod) { - auto &coordtype = expression_type(coord); string coord_filler; - for (uint32_t size = coordtype.vecsize; size < 3; ++size) + for (uint32_t size = coord_components; size < 3; ++size) { coord_filler += ", 0.0"; } @@ -2737,9 +2737,8 @@ void CompilerHLSL::emit_texture_op(const Instruction &i) if (hlsl_options.shader_model < 40 && bias) { - auto &coordtype = expression_type(coord); string coord_filler; - for (uint32_t size = coordtype.vecsize; size < 3; ++size) + for (uint32_t size = coord_components; size < 3; ++size) { coord_filler += ", 0.0"; } @@ -2748,10 +2747,9 @@ void CompilerHLSL::emit_texture_op(const Instruction &i) if (op == OpImageFetch) { - auto &coordtype = expression_type(coord); if (imgtype.image.dim != DimBuffer && !imgtype.image.ms) coord_expr = - join("int", coordtype.vecsize + 1, "(", coord_expr, ", ", lod ? to_expression(lod) : string("0"), ")"); + join("int", coord_components + 1, "(", coord_expr, ", ", lod ? to_expression(lod) : string("0"), ")"); } else expr += ", "; @@ -3030,7 +3028,7 @@ string CompilerHLSL::bitcast_glsl_op(const SPIRType &out_type, const SPIRType &i if (!requires_explicit_fp16_packing) { requires_explicit_fp16_packing = true; - force_recompile = true; + force_recompile(); } return "SPIRV_Cross_unpackFloat2x16"; } @@ -3039,7 +3037,7 @@ string CompilerHLSL::bitcast_glsl_op(const SPIRType &out_type, const SPIRType &i if (!requires_explicit_fp16_packing) { requires_explicit_fp16_packing = true; - force_recompile = true; + force_recompile(); } return "SPIRV_Cross_packFloat2x16"; } @@ -3101,7 +3099,7 @@ void CompilerHLSL::emit_glsl_op(uint32_t result_type, uint32_t id, uint32_t eop, if (!requires_fp16_packing) { requires_fp16_packing = true; - force_recompile = true; + force_recompile(); } emit_unary_func_op(result_type, id, args[0], "SPIRV_Cross_packHalf2x16"); break; @@ -3110,7 +3108,7 @@ void CompilerHLSL::emit_glsl_op(uint32_t result_type, uint32_t id, uint32_t eop, if (!requires_fp16_packing) { requires_fp16_packing = true; - force_recompile = true; + force_recompile(); } emit_unary_func_op(result_type, id, args[0], "SPIRV_Cross_unpackHalf2x16"); break; @@ -3119,7 +3117,7 @@ void CompilerHLSL::emit_glsl_op(uint32_t result_type, uint32_t id, uint32_t eop, if (!requires_snorm8_packing) { requires_snorm8_packing = true; - force_recompile = true; + force_recompile(); } emit_unary_func_op(result_type, id, args[0], "SPIRV_Cross_packSnorm4x8"); break; @@ -3128,7 +3126,7 @@ void CompilerHLSL::emit_glsl_op(uint32_t result_type, uint32_t id, uint32_t eop, if (!requires_snorm8_packing) { requires_snorm8_packing = true; - force_recompile = true; + force_recompile(); } emit_unary_func_op(result_type, id, args[0], "SPIRV_Cross_unpackSnorm4x8"); break; @@ -3137,7 +3135,7 @@ void CompilerHLSL::emit_glsl_op(uint32_t result_type, uint32_t id, uint32_t eop, if (!requires_unorm8_packing) { requires_unorm8_packing = true; - force_recompile = true; + force_recompile(); } emit_unary_func_op(result_type, id, args[0], "SPIRV_Cross_packUnorm4x8"); break; @@ -3146,7 +3144,7 @@ void CompilerHLSL::emit_glsl_op(uint32_t result_type, uint32_t id, uint32_t eop, if (!requires_unorm8_packing) { requires_unorm8_packing = true; - force_recompile = true; + force_recompile(); } emit_unary_func_op(result_type, id, args[0], "SPIRV_Cross_unpackUnorm4x8"); break; @@ -3155,7 +3153,7 @@ void CompilerHLSL::emit_glsl_op(uint32_t result_type, uint32_t id, uint32_t eop, if (!requires_snorm16_packing) { requires_snorm16_packing = true; - force_recompile = true; + force_recompile(); } emit_unary_func_op(result_type, id, args[0], "SPIRV_Cross_packSnorm2x16"); break; @@ -3164,7 +3162,7 @@ void CompilerHLSL::emit_glsl_op(uint32_t result_type, uint32_t id, uint32_t eop, if (!requires_snorm16_packing) { requires_snorm16_packing = true; - force_recompile = true; + force_recompile(); } emit_unary_func_op(result_type, id, args[0], "SPIRV_Cross_unpackSnorm2x16"); break; @@ -3173,7 +3171,7 @@ void CompilerHLSL::emit_glsl_op(uint32_t result_type, uint32_t id, uint32_t eop, if (!requires_unorm16_packing) { requires_unorm16_packing = true; - force_recompile = true; + force_recompile(); } emit_unary_func_op(result_type, id, args[0], "SPIRV_Cross_packUnorm2x16"); break; @@ -3182,7 +3180,7 @@ void CompilerHLSL::emit_glsl_op(uint32_t result_type, uint32_t id, uint32_t eop, if (!requires_unorm16_packing) { requires_unorm16_packing = true; - force_recompile = true; + force_recompile(); } emit_unary_func_op(result_type, id, args[0], "SPIRV_Cross_unpackUnorm2x16"); break; @@ -3211,7 +3209,7 @@ void CompilerHLSL::emit_glsl_op(uint32_t result_type, uint32_t id, uint32_t eop, if (!requires_inverse_2x2) { requires_inverse_2x2 = true; - force_recompile = true; + force_recompile(); } } else if (type.vecsize == 3 && type.columns == 3) @@ -3219,7 +3217,7 @@ void CompilerHLSL::emit_glsl_op(uint32_t result_type, uint32_t id, uint32_t eop, if (!requires_inverse_3x3) { requires_inverse_3x3 = true; - force_recompile = true; + force_recompile(); } } else if (type.vecsize == 4 && type.columns == 4) @@ -3227,7 +3225,7 @@ void CompilerHLSL::emit_glsl_op(uint32_t result_type, uint32_t id, uint32_t eop, if (!requires_inverse_4x4) { requires_inverse_4x4 = true; - force_recompile = true; + force_recompile(); } } emit_unary_func_op(result_type, id, args[0], "SPIRV_Cross_Inverse"); @@ -3949,7 +3947,7 @@ void CompilerHLSL::emit_instruction(const Instruction &instruction) if (!requires_op_fmod) { requires_op_fmod = true; - force_recompile = true; + force_recompile(); } CompilerGLSL::emit_instruction(instruction); break; @@ -4464,7 +4462,7 @@ void CompilerHLSL::emit_instruction(const Instruction &instruction) if (!requires_bitfield_insert) { requires_bitfield_insert = true; - force_recompile = true; + force_recompile(); } auto expr = join("SPIRV_Cross_bitfieldInsert(", to_expression(ops[2]), ", ", to_expression(ops[3]), ", ", @@ -4485,7 +4483,7 @@ void CompilerHLSL::emit_instruction(const Instruction &instruction) if (!requires_bitfield_extract) { requires_bitfield_extract = true; - force_recompile = true; + force_recompile(); } if (opcode == OpBitFieldSExtract) @@ -4562,7 +4560,7 @@ void CompilerHLSL::require_texture_query_variant(const SPIRType &type) uint64_t mask = 1ull << bit; if ((required_textureSizeVariants & mask) == 0) { - force_recompile = true; + force_recompile(); required_textureSizeVariants |= mask; } } @@ -4675,7 +4673,7 @@ string CompilerHLSL::compile() emit_hlsl_entry_point(); pass_count++; - } while (force_recompile); + } while (is_forcing_recompilation()); // Entry point in HLSL is always main() for the time being. get_entry_point().name = "main"; diff --git a/3rdparty/spirv-cross/spirv_msl.cpp b/3rdparty/spirv-cross/spirv_msl.cpp index 58c198dcc..72fb872a5 100644 --- a/3rdparty/spirv-cross/spirv_msl.cpp +++ b/3rdparty/spirv-cross/spirv_msl.cpp @@ -668,7 +668,7 @@ string CompilerMSL::compile() emit_function(get(ir.default_entry_point), Bitset()); pass_count++; - } while (force_recompile); + } while (is_forcing_recompilation()); return buffer->str(); } @@ -2378,14 +2378,14 @@ void CompilerMSL::add_pragma_line(const string &line) { auto rslt = pragma_lines.insert(line); if (rslt.second) - force_recompile = true; + force_recompile(); } void CompilerMSL::add_typedef_line(const string &line) { auto rslt = typedef_lines.insert(line); if (rslt.second) - force_recompile = true; + force_recompile(); } // Emits any needed custom function bodies. @@ -3569,7 +3569,7 @@ void CompilerMSL::emit_instruction(const Instruction &instruction) if (p_var && has_decoration(p_var->self, DecorationNonReadable)) { unset_decoration(p_var->self, DecorationNonReadable); - force_recompile = true; + force_recompile(); } } @@ -3595,7 +3595,7 @@ void CompilerMSL::emit_instruction(const Instruction &instruction) if (p_var && has_decoration(p_var->self, DecorationNonWritable)) { unset_decoration(p_var->self, DecorationNonWritable); - force_recompile = true; + force_recompile(); } bool forward = false; @@ -4043,13 +4043,17 @@ void CompilerMSL::emit_atomic_func_op(uint32_t result_type, uint32_t result_id, exp += get_memory_order(mem_order_2); exp += ")"; - // MSL only supports the weak atomic compare exchange, - // so emit a CAS loop here. + // MSL only supports the weak atomic compare exchange, so emit a CAS loop here. + // The MSL function returns false if the atomic write fails OR the comparison test fails, + // so we must validate that it wasn't the comparison test that failed before continuing + // the CAS loop, otherwise it will loop infinitely, with the comparison test always failing. + // The function updates the comparitor value from the memory value, so the additional + // comparison test evaluates the memory value against the expected value. statement(variable_decl(type, to_name(result_id)), ";"); statement("do"); begin_scope(); statement(to_name(result_id), " = ", to_expression(op1), ";"); - end_scope_decl(join("while (!", exp, ")")); + end_scope_decl(join("while (!", exp, " && ", to_name(result_id), " == ", to_enclosed_expression(op1), ")")); set(result_id, to_name(result_id), result_type, true); } else @@ -4792,7 +4796,7 @@ string CompilerMSL::to_func_call_arg(uint32_t id) auto itr = find(begin(constants), end(constants), id); if (itr == end(constants)) { - force_recompile = true; + force_recompile(); constants.push_back(id); } } @@ -4940,7 +4944,7 @@ void CompilerMSL::add_convert_row_major_matrix_function(uint32_t cols, uint32_t if (rslt.second) { add_pragma_line("#pragma clang diagnostic ignored \"-Wmissing-prototypes\""); - force_recompile = true; + force_recompile(); } }