Updated spirv-cross.

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<A<int> > instead of A<A<int>> 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

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

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

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."

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];

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;
 }
 

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);
 }
 

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;
+}
+

3rdparty/spirv-cross/reference/shaders-hlsl-no-opt/asm/frag/image-fetch-uint-coord.asm.frag

@@ -0,0 +1,28 @@
+Texture2D<float4> 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;
+}

3rdparty/spirv-cross/reference/shaders-msl-no-opt/asm/frag/image-fetch-uint-coord.asm.frag

@@ -0,0 +1,22 @@
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+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<float> 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;
+}
+

3rdparty/spirv-cross/reference/shaders-msl-no-opt/asm/frag/image-type-normal-comparison-usage.asm.frag

@@ -0,0 +1,31 @@
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+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<float> 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;
+}
+

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);
 }
 

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));
+}
+

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;
+}
+

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

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

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

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

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

3rdparty/spirv-cross/spirv_cpp.cpp

@@ -342,7 +342,7 @@ string CompilerCPP::compile()
 		emit_function(get<SPIRFunction>(ir.default_entry_point), Bitset());
 
 		pass_count++;
-	} while (force_recompile);
+	} while (is_forcing_recompilation());
 
 	// Match opening scope of emit_header().
 	end_scope_decl();

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;
+}

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;
 

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<SPIRFunction>(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<SPIRCombinedImageSampler>(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<SPIRType>(result_type), 1, expr);
+		bool image_is_depth = false;
+		const auto *combined = maybe_get<SPIRCombinedImageSampler>(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<SPIRType>(image_variable->basetype), image_variable->self))
+			image_is_depth = true;
+
+		if (image_is_depth)
+			expr = remap_swizzle(get<SPIRType>(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<SPIRVariable>(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<SPIRBlock>(continue_block.true_block),
 			                                       get<SPIRBlock>(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<SPIRBlock>(block.continue_block).complex_continue = true;
-				force_recompile = true;
+				force_recompile();
 			}
 
 			// Might have to invert the do-while test here.

3rdparty/spirv-cross/spirv_glsl.hpp

@@ -280,7 +280,7 @@ protected:
 	template <typename... Ts>
 	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>(ts)...));
+			statement_count++;
+		}
 		else
 		{
 			for (uint32_t i = 0; i < indent; i++)

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";

3rdparty/spirv-cross/spirv_msl.cpp

@@ -668,7 +668,7 @@ string CompilerMSL::compile()
 		emit_function(get<SPIRFunction>(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<SPIRExpression>(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();
 	}
 }