Update SPIRV-Cross

This also pulls in the addition of the emit_uniform_buffer_as_plain_uniforms
flag for the GLSL generator, which is something we want to use.

Change-Id: Iffc6c40265c72930665587d26d8c4b9638698a01
Reviewed-by: Laszlo Agocs <laszlo.agocs@qt.io>

26 files changed, 2315 insertions, 634 deletions

diff --git a/src/3rdparty/SPIRV-Cross/qt_attribution.json b/src/3rdparty/SPIRV-Cross/qt_attribution.json
index 46ba2b9..70df546 100644
--- a/src/3rdparty/SPIRV-Cross/qt_attribution.json
+++ b/src/3rdparty/SPIRV-Cross/qt_attribution.json
@@ -7,10 +7,10 @@
         "QtUsage": "Shader code generation.",
 
         "Homepage": "https://github.com/KhronosGroup/SPIRV-Cross",
-        "Version": "3fa09f5677c7a62c71a1c25fd09c1d1c4842d922",
+        "Version": "f647e655d489a7699305ada30cda808a7dac079f",
         "License": "Apache License 2.0",
         "LicenseId": "Apache-2.0",
         "LicenseFile": "LICENSE",
-        "Copyright": "Copyright 2016-2018 ARM Limited"
+        "Copyright": "Copyright 2015-2019 ARM Limited"
     }
 ]
diff --git a/src/3rdparty/SPIRV-Cross/spirv_cfg.cpp b/src/3rdparty/SPIRV-Cross/spirv_cfg.cpp
index 4ca9ef5..2f3cf25 100644
--- a/src/3rdparty/SPIRV-Cross/spirv_cfg.cpp
+++ b/src/3rdparty/SPIRV-Cross/spirv_cfg.cpp
@@ -21,7 +21,7 @@
 
 using namespace std;
 
-namespace spirv_cross
+namespace SPIRV_CROSS_NAMESPACE
 {
 CFG::CFG(Compiler &compiler_, const SPIRFunction &func_)
     : compiler(compiler_)
@@ -143,7 +143,7 @@ void CFG::build_post_order_visit_order()
 
 void CFG::add_branch(uint32_t from, uint32_t to)
 {
-	const auto add_unique = [](vector<uint32_t> &l, uint32_t value) {
+	const auto add_unique = [](SmallVector<uint32_t> &l, uint32_t value) {
 		auto itr = find(begin(l), end(l), value);
 		if (itr == end(l))
 			l.push_back(value);
@@ -223,4 +223,4 @@ void DominatorBuilder::lift_continue_block_dominator()
 	if (back_edge_dominator)
 		dominator = cfg.get_function().entry_block;
 }
-} // namespace spirv_cross
+} // namespace SPIRV_CROSS_NAMESPACE
diff --git a/src/3rdparty/SPIRV-Cross/spirv_cfg.hpp b/src/3rdparty/SPIRV-Cross/spirv_cfg.hpp
index 5e89320..be10371 100644
--- a/src/3rdparty/SPIRV-Cross/spirv_cfg.hpp
+++ b/src/3rdparty/SPIRV-Cross/spirv_cfg.hpp
@@ -20,7 +20,7 @@
 #include "spirv_common.hpp"
 #include <assert.h>
 
-namespace spirv_cross
+namespace SPIRV_CROSS_NAMESPACE
 {
 class Compiler;
 class CFG
@@ -63,7 +63,7 @@ public:
 
 	uint32_t find_common_dominator(uint32_t a, uint32_t b) const;
 
-	const std::vector<uint32_t> &get_preceding_edges(uint32_t block) const
+	const SmallVector<uint32_t> &get_preceding_edges(uint32_t block) const
 	{
 		auto itr = preceding_edges.find(block);
 		if (itr != std::end(preceding_edges))
@@ -72,7 +72,7 @@ public:
 			return empty_vector;
 	}
 
-	const std::vector<uint32_t> &get_succeeding_edges(uint32_t block) const
+	const SmallVector<uint32_t> &get_succeeding_edges(uint32_t block) const
 	{
 		auto itr = succeeding_edges.find(block);
 		if (itr != std::end(succeeding_edges))
@@ -111,12 +111,12 @@ private:
 
 	Compiler &compiler;
 	const SPIRFunction &func;
-	std::unordered_map<uint32_t, std::vector<uint32_t>> preceding_edges;
-	std::unordered_map<uint32_t, std::vector<uint32_t>> succeeding_edges;
+	std::unordered_map<uint32_t, SmallVector<uint32_t>> preceding_edges;
+	std::unordered_map<uint32_t, SmallVector<uint32_t>> succeeding_edges;
 	std::unordered_map<uint32_t, uint32_t> immediate_dominators;
 	std::unordered_map<uint32_t, VisitOrder> visit_order;
-	std::vector<uint32_t> post_order;
-	std::vector<uint32_t> empty_vector;
+	SmallVector<uint32_t> post_order;
+	SmallVector<uint32_t> empty_vector;
 
 	void add_branch(uint32_t from, uint32_t to);
 	void build_post_order_visit_order();
@@ -144,6 +144,6 @@ private:
 	const CFG &cfg;
 	uint32_t dominator = 0;
 };
-} // namespace spirv_cross
+} // namespace SPIRV_CROSS_NAMESPACE
 
 #endif
diff --git a/src/3rdparty/SPIRV-Cross/spirv_common.hpp b/src/3rdparty/SPIRV-Cross/spirv_common.hpp
index dcd27af..0cf1f56 100644
--- a/src/3rdparty/SPIRV-Cross/spirv_common.hpp
+++ b/src/3rdparty/SPIRV-Cross/spirv_common.hpp
@@ -18,87 +18,37 @@
 #define SPIRV_CROSS_COMMON_HPP
 
 #include "spirv.hpp"
-
-#include <algorithm>
-#include <cstdio>
-#include <cstring>
-#include <functional>
-#include <memory>
-#include <sstream>
-#include <stack>
-#include <stdexcept>
-#include <stdint.h>
-#include <string>
-#include <type_traits>
-#include <unordered_map>
-#include <unordered_set>
-#include <utility>
-#include <vector>
-
-namespace spirv_cross
-{
-
-#ifdef SPIRV_CROSS_EXCEPTIONS_TO_ASSERTIONS
-#ifndef _MSC_VER
-[[noreturn]]
-#endif
-inline void
-report_and_abort(const std::string &msg)
-{
-#ifdef NDEBUG
-	(void)msg;
+#include "spirv_cross_containers.hpp"
+#include "spirv_cross_error_handling.hpp"
+
+// A bit crude, but allows projects which embed SPIRV-Cross statically to
+// effectively hide all the symbols from other projects.
+// There is a case where we have:
+// - Project A links against SPIRV-Cross statically.
+// - Project A links against Project B statically.
+// - Project B links against SPIRV-Cross statically (might be a different version).
+// This leads to a conflict with extremely bizarre results.
+// By overriding the namespace in one of the project builds, we can work around this.
+// If SPIRV-Cross is embedded in dynamic libraries,
+// prefer using -fvisibility=hidden on GCC/Clang instead.
+#ifdef SPIRV_CROSS_NAMESPACE_OVERRIDE
+#define SPIRV_CROSS_NAMESPACE SPIRV_CROSS_NAMESPACE_OVERRIDE
 #else
-	fprintf(stderr, "There was a compiler error: %s\n", msg.c_str());
+#define SPIRV_CROSS_NAMESPACE spirv_cross
 #endif
-	fflush(stderr);
-	abort();
-}
 
-#define SPIRV_CROSS_THROW(x) report_and_abort(x)
-#else
-class CompilerError : public std::runtime_error
+namespace SPIRV_CROSS_NAMESPACE
 {
-public:
-	explicit CompilerError(const std::string &str)
-	    : std::runtime_error(str)
-	{
-	}
-};
-
-#define SPIRV_CROSS_THROW(x) throw CompilerError(x)
-#endif
-
-//#define SPIRV_CROSS_COPY_CONSTRUCTOR_SANITIZE
-
-// MSVC 2013 does not have noexcept. We need this for Variant to get move constructor to work correctly
-// instead of copy constructor.
-// MSVC 2013 ignores that move constructors cannot throw in std::vector, so just don't define it.
-#if defined(_MSC_VER) && _MSC_VER < 1900
-#define SPIRV_CROSS_NOEXCEPT
-#else
-#define SPIRV_CROSS_NOEXCEPT noexcept
-#endif
-
-#if __cplusplus >= 201402l
-#define SPIRV_CROSS_DEPRECATED(reason) [[deprecated(reason)]]
-#elif defined(__GNUC__)
-#define SPIRV_CROSS_DEPRECATED(reason) __attribute__((deprecated))
-#elif defined(_MSC_VER)
-#define SPIRV_CROSS_DEPRECATED(reason) __declspec(deprecated(reason))
-#else
-#define SPIRV_CROSS_DEPRECATED(reason)
-#endif
-
 namespace inner
 {
 template <typename T>
-void join_helper(std::ostringstream &stream, T &&t)
+void join_helper(StringStream<> &stream, T &&t)
 {
 	stream << std::forward<T>(t);
 }
 
 template <typename T, typename... Ts>
-void join_helper(std::ostringstream &stream, T &&t, Ts &&... ts)
+void join_helper(StringStream<> &stream, T &&t, Ts &&... ts)
 {
 	stream << std::forward<T>(t);
 	join_helper(stream, std::forward<Ts>(ts)...);
@@ -201,7 +151,7 @@ public:
 
 		// Need to enforce an order here for reproducible results,
 		// but hitting this path should happen extremely rarely, so having this slow path is fine.
-		std::vector<uint32_t> bits;
+		SmallVector<uint32_t> bits;
 		bits.reserve(higher.size());
 		for (auto &v : higher)
 			bits.push_back(v);
@@ -228,21 +178,21 @@ private:
 template <typename... Ts>
 std::string join(Ts &&... ts)
 {
-	std::ostringstream stream;
+	StringStream<> stream;
 	inner::join_helper(stream, std::forward<Ts>(ts)...);
 	return stream.str();
 }
 
-inline std::string merge(const std::vector<std::string> &list)
+inline std::string merge(const SmallVector<std::string> &list)
 {
-	std::string s;
+	StringStream<> stream;
 	for (auto &elem : list)
 	{
-		s += elem;
+		stream << elem;
 		if (&elem != &list.back())
-			s += ", ";
+			stream << ", ";
 	}
-	return s;
+	return stream.str();
 }
 
 // Make sure we don't accidentally call this with float or doubles with SFINAE.
@@ -324,15 +274,14 @@ struct Instruction
 struct IVariant
 {
 	virtual ~IVariant() = default;
-	virtual std::unique_ptr<IVariant> clone() = 0;
-
+	virtual IVariant *clone(ObjectPoolBase *pool) = 0;
 	uint32_t self = 0;
 };
 
-#define SPIRV_CROSS_DECLARE_CLONE(T)                    \
-	std::unique_ptr<IVariant> clone() override          \
-	{                                                   \
-		return std::unique_ptr<IVariant>(new T(*this)); \
+#define SPIRV_CROSS_DECLARE_CLONE(T)                                \
+	IVariant *clone(ObjectPoolBase *pool) override                  \
+	{                                                               \
+		return static_cast<ObjectPool<T> *>(pool)->allocate(*this); \
 	}
 
 enum Types
@@ -405,7 +354,7 @@ struct SPIRConstantOp : IVariant
 	}
 
 	spv::Op opcode;
-	std::vector<uint32_t> arguments;
+	SmallVector<uint32_t> arguments;
 	uint32_t basetype;
 
 	SPIRV_CROSS_DECLARE_CLONE(SPIRConstantOp)
@@ -453,14 +402,14 @@ struct SPIRType : IVariant
 	uint32_t columns = 1;
 
 	// Arrays, support array of arrays by having a vector of array sizes.
-	std::vector<uint32_t> array;
+	SmallVector<uint32_t> array;
 
 	// Array elements can be either specialization constants or specialization ops.
 	// This array determines how to interpret the array size.
 	// If an element is true, the element is a literal,
 	// otherwise, it's an expression, which must be resolved on demand.
 	// The actual size is not really known until runtime.
-	std::vector<bool> array_size_literal;
+	SmallVector<bool> array_size_literal;
 
 	// Pointers
 	// Keep track of how many pointer layers we have.
@@ -469,7 +418,7 @@ struct SPIRType : IVariant
 
 	spv::StorageClass storage = spv::StorageClassGeneric;
 
-	std::vector<uint32_t> member_types;
+	SmallVector<uint32_t> member_types;
 
 	struct ImageType
 	{
@@ -540,7 +489,7 @@ struct SPIREntryPoint
 	uint32_t self = 0;
 	std::string name;
 	std::string orig_name;
-	std::vector<uint32_t> interface_variables;
+	SmallVector<uint32_t> interface_variables;
 
 	Bitset flags;
 	struct
@@ -594,11 +543,11 @@ struct SPIRExpression : IVariant
 	bool access_chain = false;
 
 	// A list of expressions which this expression depends on.
-	std::vector<uint32_t> expression_dependencies;
+	SmallVector<uint32_t> expression_dependencies;
 
 	// By reading this expression, we implicitly read these expressions as well.
 	// Used by access chain Store and Load since we read multiple expressions in this case.
-	std::vector<uint32_t> implied_read_expressions;
+	SmallVector<uint32_t> implied_read_expressions;
 
 	SPIRV_CROSS_DECLARE_CLONE(SPIRExpression)
 };
@@ -616,7 +565,7 @@ struct SPIRFunctionPrototype : IVariant
 	}
 
 	uint32_t return_type;
-	std::vector<uint32_t> parameter_types;
+	SmallVector<uint32_t> parameter_types;
 
 	SPIRV_CROSS_DECLARE_CLONE(SPIRFunctionPrototype)
 };
@@ -700,7 +649,7 @@ struct SPIRBlock : IVariant
 	uint32_t false_block = 0;
 	uint32_t default_block = 0;
 
-	std::vector<Instruction> ops;
+	SmallVector<Instruction> ops;
 
 	struct Phi
 	{
@@ -710,22 +659,22 @@ struct SPIRBlock : IVariant
 	};
 
 	// Before entering this block flush out local variables to magical "phi" variables.
-	std::vector<Phi> phi_variables;
+	SmallVector<Phi> phi_variables;
 
 	// Declare these temporaries before beginning the block.
 	// Used for handling complex continue blocks which have side effects.
-	std::vector<std::pair<uint32_t, uint32_t>> declare_temporary;
+	SmallVector<std::pair<uint32_t, uint32_t>> declare_temporary;
 
 	// Declare these temporaries, but only conditionally if this block turns out to be
 	// a complex loop header.
-	std::vector<std::pair<uint32_t, uint32_t>> potential_declare_temporary;
+	SmallVector<std::pair<uint32_t, uint32_t>> potential_declare_temporary;
 
 	struct Case
 	{
 		uint32_t value;
 		uint32_t block;
 	};
-	std::vector<Case> cases;
+	SmallVector<Case> cases;
 
 	// If we have tried to optimize code for this block but failed,
 	// keep track of this.
@@ -743,17 +692,17 @@ struct SPIRBlock : IVariant
 
 	// All access to these variables are dominated by this block,
 	// so before branching anywhere we need to make sure that we declare these variables.
-	std::vector<uint32_t> dominated_variables;
+	SmallVector<uint32_t> dominated_variables;
 
 	// These are variables which should be declared in a for loop header, if we
 	// fail to use a classic for-loop,
 	// we remove these variables, and fall back to regular variables outside the loop.
-	std::vector<uint32_t> loop_variables;
+	SmallVector<uint32_t> loop_variables;
 
 	// Some expressions are control-flow dependent, i.e. any instruction which relies on derivatives or
 	// sub-group-like operations.
 	// Make sure that we only use these expressions in the original block.
-	std::vector<uint32_t> invalidate_expressions;
+	SmallVector<uint32_t> invalidate_expressions;
 
 	SPIRV_CROSS_DECLARE_CLONE(SPIRBlock)
 };
@@ -806,16 +755,16 @@ struct SPIRFunction : IVariant
 
 	uint32_t return_type;
 	uint32_t function_type;
-	std::vector<Parameter> arguments;
+	SmallVector<Parameter> arguments;
 
 	// Can be used by backends to add magic arguments.
 	// Currently used by combined image/sampler implementation.
 
-	std::vector<Parameter> shadow_arguments;
-	std::vector<uint32_t> local_variables;
+	SmallVector<Parameter> shadow_arguments;
+	SmallVector<uint32_t> local_variables;
 	uint32_t entry_block = 0;
-	std::vector<uint32_t> blocks;
-	std::vector<CombinedImageSamplerParameter> combined_parameters;
+	SmallVector<uint32_t> blocks;
+	SmallVector<CombinedImageSamplerParameter> combined_parameters;
 
 	void add_local_variable(uint32_t id)
 	{
@@ -831,17 +780,19 @@ struct SPIRFunction : IVariant
 	// Hooks to be run when the function returns.
 	// Mostly used for lowering internal data structures onto flattened structures.
 	// Need to defer this, because they might rely on things which change during compilation.
-	std::vector<std::function<void()>> fixup_hooks_out;
+	// Intentionally not a small vector, this one is rare, and std::function can be large.
+	Vector<std::function<void()>> fixup_hooks_out;
 
 	// Hooks to be run when the function begins.
 	// Mostly used for populating internal data structures from flattened structures.
 	// Need to defer this, because they might rely on things which change during compilation.
-	std::vector<std::function<void()>> fixup_hooks_in;
+	// Intentionally not a small vector, this one is rare, and std::function can be large.
+	Vector<std::function<void()>> fixup_hooks_in;
 
 	// On function entry, make sure to copy a constant array into thread addr space to work around
 	// the case where we are passing a constant array by value to a function on backends which do not
 	// consider arrays value types.
-	std::vector<uint32_t> constant_arrays_needed_on_stack;
+	SmallVector<uint32_t> constant_arrays_needed_on_stack;
 
 	bool active = false;
 	bool flush_undeclared = true;
@@ -885,7 +836,7 @@ struct SPIRAccessChain : IVariant
 
 	// By reading this expression, we implicitly read these expressions as well.
 	// Used by access chain Store and Load since we read multiple expressions in this case.
-	std::vector<uint32_t> implied_read_expressions;
+	SmallVector<uint32_t> implied_read_expressions;
 
 	SPIRV_CROSS_DECLARE_CLONE(SPIRAccessChain)
 };
@@ -912,7 +863,7 @@ struct SPIRVariable : IVariant
 	uint32_t initializer = 0;
 	uint32_t basevariable = 0;
 
-	std::vector<uint32_t> dereference_chain;
+	SmallVector<uint32_t> dereference_chain;
 	bool compat_builtin = false;
 
 	// If a variable is shadowed, we only statically assign to it
@@ -923,7 +874,7 @@ struct SPIRVariable : IVariant
 	uint32_t static_expression = 0;
 
 	// Temporaries which can remain forwarded as long as this variable is not modified.
-	std::vector<uint32_t> dependees;
+	SmallVector<uint32_t> dependees;
 	bool forwardable = true;
 
 	bool deferred_declaration = false;
@@ -1162,7 +1113,7 @@ struct SPIRConstant : IVariant
 	    : constant_type(constant_type_)
 	    , specialization(specialized)
 	{
-		subconstants.insert(end(subconstants), elements, elements + num_elements);
+		subconstants.insert(std::end(subconstants), elements, elements + num_elements);
 		specialization = specialized;
 	}
 
@@ -1231,7 +1182,7 @@ struct SPIRConstant : IVariant
 	bool is_used_as_lut = false;
 
 	// For composites which are constant arrays, etc.
-	std::vector<uint32_t> subconstants;
+	SmallVector<uint32_t> subconstants;
 
 	// Non-Vulkan GLSL, HLSL and sometimes MSL emits defines for each specialization constant,
 	// and uses them to initialize the constant. This allows the user
@@ -1242,11 +1193,25 @@ struct SPIRConstant : IVariant
 	SPIRV_CROSS_DECLARE_CLONE(SPIRConstant)
 };
 
+// Variants have a very specific allocation scheme.
+struct ObjectPoolGroup
+{
+	std::unique_ptr<ObjectPoolBase> pools[TypeCount];
+};
+
 class Variant
 {
 public:
-	// MSVC 2013 workaround, we shouldn't need these constructors.
-	Variant() = default;
+	explicit Variant(ObjectPoolGroup *group_)
+	    : group(group_)
+	{
+	}
+
+	~Variant()
+	{
+		if (holder)
+			group->pools[type]->free_opaque(holder);
+	}
 
 	// Marking custom move constructor as noexcept is important.
 	Variant(Variant &&other) SPIRV_CROSS_NOEXCEPT
@@ -1254,19 +1219,23 @@ public:
 		*this = std::move(other);
 	}
 
-	Variant(const Variant &variant)
-	{
-		*this = variant;
-	}
+	// We cannot copy from other variant without our own pool group.
+	// Have to explicitly copy.
+	Variant(const Variant &variant) = delete;
 
 	// Marking custom move constructor as noexcept is important.
 	Variant &operator=(Variant &&other) SPIRV_CROSS_NOEXCEPT
 	{
 		if (this != &other)
 		{
-			holder = std::move(other.holder);
+			if (holder)
+				group->pools[type]->free_opaque(holder);
+			holder = other.holder;
+			group = other.group;
 			type = other.type;
 			allow_type_rewrite = other.allow_type_rewrite;
+
+			other.holder = nullptr;
 			other.type = TypeNone;
 		}
 		return *this;
@@ -1277,29 +1246,52 @@ public:
 	// This should never happen.
 	Variant &operator=(const Variant &other)
 	{
+//#define SPIRV_CROSS_COPY_CONSTRUCTOR_SANITIZE
 #ifdef SPIRV_CROSS_COPY_CONSTRUCTOR_SANITIZE
 		abort();
 #endif
 		if (this != &other)
 		{
-			holder.reset();
+			if (holder)
+				group->pools[type]->free_opaque(holder);
+
 			if (other.holder)
-				holder = other.holder->clone();
+				holder = other.holder->clone(group->pools[other.type].get());
+			else
+				holder = nullptr;
+
 			type = other.type;
 			allow_type_rewrite = other.allow_type_rewrite;
 		}
 		return *this;
 	}
 
-	void set(std::unique_ptr<IVariant> val, Types new_type)
+	void set(IVariant *val, Types new_type)
 	{
-		holder = std::move(val);
+		if (holder)
+			group->pools[type]->free_opaque(holder);
+		holder = nullptr;
+
 		if (!allow_type_rewrite && type != TypeNone && type != new_type)
+		{
+			if (val)
+				group->pools[new_type]->free_opaque(val);
 			SPIRV_CROSS_THROW("Overwriting a variant with new type.");
+		}
+
+		holder = val;
 		type = new_type;
 		allow_type_rewrite = false;
 	}
 
+	template <typename T, typename... Ts>
+	T *allocate_and_set(Types new_type, Ts &&... ts)
+	{
+		T *val = static_cast<ObjectPool<T> &>(*group->pools[new_type]).allocate(std::forward<Ts>(ts)...);
+		set(val, new_type);
+		return val;
+	}
+
 	template <typename T>
 	T &get()
 	{
@@ -1307,7 +1299,7 @@ public:
 			SPIRV_CROSS_THROW("nullptr");
 		if (static_cast<Types>(T::type) != type)
 			SPIRV_CROSS_THROW("Bad cast");
-		return *static_cast<T *>(holder.get());
+		return *static_cast<T *>(holder);
 	}
 
 	template <typename T>
@@ -1317,7 +1309,7 @@ public:
 			SPIRV_CROSS_THROW("nullptr");
 		if (static_cast<Types>(T::type) != type)
 			SPIRV_CROSS_THROW("Bad cast");
-		return *static_cast<const T *>(holder.get());
+		return *static_cast<const T *>(holder);
 	}
 
 	Types get_type() const
@@ -1337,7 +1329,9 @@ public:
 
 	void reset()
 	{
-		holder.reset();
+		if (holder)
+			group->pools[type]->free_opaque(holder);
+		holder = nullptr;
 		type = TypeNone;
 	}
 
@@ -1347,7 +1341,8 @@ public:
 	}
 
 private:
-	std::unique_ptr<IVariant> holder;
+	ObjectPoolGroup *group = nullptr;
+	IVariant *holder = nullptr;
 	Types type = TypeNone;
 	bool allow_type_rewrite = false;
 };
@@ -1367,9 +1362,7 @@ const T &variant_get(const Variant &var)
 template <typename T, typename... P>
 T &variant_set(Variant &var, P &&... args)
 {
-	auto uptr = std::unique_ptr<T>(new T(std::forward<P>(args)...));
-	auto ptr = uptr.get();
-	var.set(std::move(uptr), static_cast<Types>(T::type));
+	auto *ptr = var.allocate_and_set<T>(static_cast<Types>(T::type), std::forward<P>(args)...);
 	return *ptr;
 }
 
@@ -1414,7 +1407,9 @@ struct Meta
 	};
 
 	Decoration decoration;
-	std::vector<Decoration> members;
+
+	// Intentionally not a SmallVector. Decoration is large and somewhat rare.
+	Vector<Decoration> members;
 
 	std::unordered_map<uint32_t, uint32_t> decoration_word_offset;
 
@@ -1513,6 +1508,6 @@ static inline bool opcode_is_sign_invariant(spv::Op opcode)
 		return false;
 	}
 }
-} // namespace spirv_cross
+} // namespace SPIRV_CROSS_NAMESPACE
 
 #endif
diff --git a/src/3rdparty/SPIRV-Cross/spirv_cpp.cpp b/src/3rdparty/SPIRV-Cross/spirv_cpp.cpp
index 1b791ee..90566c1 100644
--- a/src/3rdparty/SPIRV-Cross/spirv_cpp.cpp
+++ b/src/3rdparty/SPIRV-Cross/spirv_cpp.cpp
@@ -17,7 +17,7 @@
 #include "spirv_cpp.hpp"
 
 using namespace spv;
-using namespace spirv_cross;
+using namespace SPIRV_CROSS_NAMESPACE;
 using namespace std;
 
 void CompilerCPP::emit_buffer_block(const SPIRVariable &var)
@@ -317,7 +317,7 @@ string CompilerCPP::compile()
 	backend.basic_uint_type = "uint32_t";
 	backend.swizzle_is_function = true;
 	backend.shared_is_implied = true;
-	backend.flexible_member_array_supported = false;
+	backend.unsized_array_supported = false;
 	backend.explicit_struct_type = true;
 	backend.use_initializer_list = true;
 
@@ -334,7 +334,7 @@ string CompilerCPP::compile()
 		reset();
 
 		// Move constructor for this type is broken on GCC 4.9 ...
-		buffer = unique_ptr<ostringstream>(new ostringstream());
+		buffer.reset();
 
 		emit_header();
 		emit_resources();
@@ -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();
@@ -355,7 +355,7 @@ string CompilerCPP::compile()
 	// Entry point in CPP is always main() for the time being.
 	get_entry_point().name = "main";
 
-	return buffer->str();
+	return buffer.str();
 }
 
 void CompilerCPP::emit_c_linkage()
diff --git a/src/3rdparty/SPIRV-Cross/spirv_cpp.hpp b/src/3rdparty/SPIRV-Cross/spirv_cpp.hpp
index bcdb669..4c20aa3 100644
--- a/src/3rdparty/SPIRV-Cross/spirv_cpp.hpp
+++ b/src/3rdparty/SPIRV-Cross/spirv_cpp.hpp
@@ -19,15 +19,14 @@
 
 #include "spirv_glsl.hpp"
 #include <utility>
-#include <vector>
 
-namespace spirv_cross
+namespace SPIRV_CROSS_NAMESPACE
 {
 class CompilerCPP : public CompilerGLSL
 {
 public:
 	explicit CompilerCPP(std::vector<uint32_t> spirv_)
-	    : CompilerGLSL(move(spirv_))
+	    : CompilerGLSL(std::move(spirv_))
 	{
 	}
 
@@ -75,13 +74,13 @@ private:
 
 	std::string argument_decl(const SPIRFunction::Parameter &arg);
 
-	std::vector<std::string> resource_registrations;
+	SmallVector<std::string> resource_registrations;
 	std::string impl_type;
 	std::string resource_type;
 	uint32_t shared_counter = 0;
 
 	std::string interface_name;
 };
-} // namespace spirv_cross
+} // namespace SPIRV_CROSS_NAMESPACE
 
 #endif
diff --git a/src/3rdparty/SPIRV-Cross/spirv_cross.cpp b/src/3rdparty/SPIRV-Cross/spirv_cross.cpp
index 7ca2fe1..6b66b74 100644
--- a/src/3rdparty/SPIRV-Cross/spirv_cross.cpp
+++ b/src/3rdparty/SPIRV-Cross/spirv_cross.cpp
@@ -24,7 +24,7 @@
 
 using namespace std;
 using namespace spv;
-using namespace spirv_cross;
+using namespace SPIRV_CROSS_NAMESPACE;
 
 Compiler::Compiler(vector<uint32_t> ir_)
 {
@@ -74,6 +74,7 @@ bool Compiler::variable_storage_is_aliased(const SPIRVariable &v)
 	            ir.meta[type.self].decoration.decoration_flags.get(DecorationBufferBlock);
 	bool image = type.basetype == SPIRType::Image;
 	bool counter = type.basetype == SPIRType::AtomicCounter;
+	bool buffer_reference = type.storage == StorageClassPhysicalStorageBufferEXT;
 
 	bool is_restrict;
 	if (ssbo)
@@ -81,7 +82,7 @@ bool Compiler::variable_storage_is_aliased(const SPIRVariable &v)
 	else
 		is_restrict = has_decoration(v.self, DecorationRestrict);
 
-	return !is_restrict && (ssbo || image || counter);
+	return !is_restrict && (ssbo || image || counter || buffer_reference);
 }
 
 bool Compiler::block_is_pure(const SPIRBlock &block)
@@ -300,21 +301,44 @@ void Compiler::register_write(uint32_t chain)
 
 	if (var)
 	{
+		bool check_argument_storage_qualifier = true;
+		auto &type = expression_type(chain);
+
 		// If our variable is in a storage class which can alias with other buffers,
 		// invalidate all variables which depend on aliased variables. And if this is a
 		// variable pointer, then invalidate all variables regardless.
 		if (get_variable_data_type(*var).pointer)
+		{
 			flush_all_active_variables();
-		if (variable_storage_is_aliased(*var))
+
+			if (type.pointer_depth == 1)
+			{
+				// We have a backing variable which is a pointer-to-pointer type.
+				// We are storing some data through a pointer acquired through that variable,
+				// but we are not writing to the value of the variable itself,
+				// i.e., we are not modifying the pointer directly.
+				// If we are storing a non-pointer type (pointer_depth == 1),
+				// we know that we are storing some unrelated data.
+				// A case here would be
+				// void foo(Foo * const *arg) {
+				//   Foo *bar = *arg;
+				//   bar->unrelated = 42;
+				// }
+				// arg, the argument is constant.
+				check_argument_storage_qualifier = false;
+			}
+		}
+
+		if (type.storage == StorageClassPhysicalStorageBufferEXT || variable_storage_is_aliased(*var))
 			flush_all_aliased_variables();
 		else if (var)
 			flush_dependees(*var);
 
 		// We tried to write to a parameter which is not marked with out qualifier, force a recompile.
-		if (var->parameter && var->parameter->write_count == 0)
+		if (check_argument_storage_qualifier && var->parameter && var->parameter->write_count == 0)
 		{
 			var->parameter->write_count++;
-			force_recompile = true;
+			force_recompile();
 		}
 	}
 	else
@@ -624,11 +648,11 @@ bool Compiler::InterfaceVariableAccessHandler::handle(Op opcode, const uint32_t
 
 		auto *var = compiler.maybe_get<SPIRVariable>(args[0]);
 		if (var && storage_class_is_interface(var->storage))
-			variables.insert(variable);
+			variables.insert(args[0]);
 
 		var = compiler.maybe_get<SPIRVariable>(args[1]);
 		if (var && storage_class_is_interface(var->storage))
-			variables.insert(variable);
+			variables.insert(args[1]);
 		break;
 	}
 
@@ -1747,7 +1771,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
@@ -1897,9 +1921,9 @@ bool Compiler::BufferAccessHandler::handle(Op opcode, const uint32_t *args, uint
 	return true;
 }
 
-std::vector<BufferRange> Compiler::get_active_buffer_ranges(uint32_t id) const
+SmallVector<BufferRange> Compiler::get_active_buffer_ranges(uint32_t id) const
 {
-	std::vector<BufferRange> ranges;
+	SmallVector<BufferRange> ranges;
 	BufferAccessHandler handler(*this, ranges, id);
 	traverse_all_reachable_opcodes(get<SPIRFunction>(ir.default_entry_point), handler);
 	return ranges;
@@ -2126,9 +2150,9 @@ void Compiler::inherit_expression_dependencies(uint32_t dst, uint32_t source_exp
 	e_deps.erase(unique(begin(e_deps), end(e_deps)), end(e_deps));
 }
 
-vector<EntryPoint> Compiler::get_entry_points_and_stages() const
+SmallVector<EntryPoint> Compiler::get_entry_points_and_stages() const
 {
-	vector<EntryPoint> entries;
+	SmallVector<EntryPoint> entries;
 	for (auto &entry : ir.entry_points)
 		entries.push_back({ entry.second.orig_name, entry.second.model });
 	return entries;
@@ -2715,9 +2739,9 @@ void Compiler::build_combined_image_samplers()
 	traverse_all_reachable_opcodes(get<SPIRFunction>(ir.default_entry_point), handler);
 }
 
-vector<SpecializationConstant> Compiler::get_specialization_constants() const
+SmallVector<SpecializationConstant> Compiler::get_specialization_constants() const
 {
-	vector<SpecializationConstant> spec_consts;
+	SmallVector<SpecializationConstant> spec_consts;
 	ir.for_each_typed_id<SPIRConstant>([&](uint32_t, const SPIRConstant &c) {
 		if (c.specialization && has_decoration(c.self, DecorationSpecId))
 			spec_consts.push_back({ c.self, get_decoration(c.self, DecorationSpecId) });
@@ -2874,6 +2898,9 @@ void Compiler::AnalyzeVariableScopeAccessHandler::set_current_block(const SPIRBl
 
 void Compiler::AnalyzeVariableScopeAccessHandler::notify_variable_access(uint32_t id, uint32_t block)
 {
+	if (id == 0)
+		return;
+
 	if (id_is_phi_variable(id))
 		accessed_variables_to_block[id].insert(block);
 	else if (id_is_potential_temporary(id))
@@ -2924,6 +2951,8 @@ bool Compiler::AnalyzeVariableScopeAccessHandler::handle(spv::Op op, const uint3
 				partial_write_variables_to_block[var->self].insert(current_block->self);
 		}
 
+		// args[0] might be an access chain we have to track use of.
+		notify_variable_access(args[0], current_block->self);
 		// Might try to store a Phi variable here.
 		notify_variable_access(args[1], current_block->self);
 		break;
@@ -2941,9 +2970,16 @@ bool Compiler::AnalyzeVariableScopeAccessHandler::handle(spv::Op op, const uint3
 		if (var)
 			accessed_variables_to_block[var->self].insert(current_block->self);
 
-		for (uint32_t i = 3; i < length; i++)
+		// args[2] might be another access chain we have to track use of.
+		for (uint32_t i = 2; i < length; i++)
 			notify_variable_access(args[i], current_block->self);
 
+		// Also keep track of the access chain pointer itself.
+		// In exceptionally rare cases, we can end up with a case where
+		// the access chain is generated in the loop body, but is consumed in continue block.
+		// This means we need complex loop workarounds, and we must detect this via CFG analysis.
+		notify_variable_access(args[1], current_block->self);
+
 		// The result of an access chain is a fixed expression and is not really considered a temporary.
 		auto &e = compiler.set<SPIRExpression>(args[1], "", args[0], true);
 		auto *backing_variable = compiler.maybe_get_backing_variable(ptr);
@@ -2951,6 +2987,7 @@ bool Compiler::AnalyzeVariableScopeAccessHandler::handle(spv::Op op, const uint3
 
 		// Other backends might use SPIRAccessChain for this later.
 		compiler.ir.ids[args[1]].set_allow_type_rewrite();
+		access_chain_expressions.insert(args[1]);
 		break;
 	}
 
@@ -2973,6 +3010,10 @@ bool Compiler::AnalyzeVariableScopeAccessHandler::handle(spv::Op op, const uint3
 				partial_write_variables_to_block[var->self].insert(current_block->self);
 		}
 
+		// args[0:1] might be access chains we have to track use of.
+		for (uint32_t i = 0; i < 2; i++)
+			notify_variable_access(args[i], current_block->self);
+
 		var = compiler.maybe_get_backing_variable(rhs);
 		if (var)
 			accessed_variables_to_block[var->self].insert(current_block->self);
@@ -2988,6 +3029,11 @@ bool Compiler::AnalyzeVariableScopeAccessHandler::handle(spv::Op op, const uint3
 		if (var)
 			accessed_variables_to_block[var->self].insert(current_block->self);
 
+		// Might be an access chain which we have to keep track of.
+		notify_variable_access(args[1], current_block->self);
+		if (access_chain_expressions.count(args[2]))
+			access_chain_expressions.insert(args[1]);
+
 		// Might try to copy a Phi variable here.
 		notify_variable_access(args[2], current_block->self);
 		break;
@@ -3004,6 +3050,9 @@ bool Compiler::AnalyzeVariableScopeAccessHandler::handle(spv::Op op, const uint3
 
 		// Loaded value is a temporary.
 		notify_variable_access(args[1], current_block->self);
+
+		// Might be an access chain we have to track use of.
+		notify_variable_access(args[2], current_block->self);
 		break;
 	}
 
@@ -3370,7 +3419,14 @@ void Compiler::analyze_variable_scope(SPIRFunction &entry, AnalyzeVariableScopeA
 			// If a temporary is used in more than one block, we might have to lift continue block
 			// access up to loop header like we did for variables.
 			if (blocks.size() != 1 && is_continue(block))
-				builder.add_block(ir.continue_block_to_loop_header[block]);
+			{
+				auto &loop_header_block = get<SPIRBlock>(ir.continue_block_to_loop_header[block]);
+				assert(loop_header_block.merge == SPIRBlock::MergeLoop);
+
+				// Only relevant if the loop is not marked as complex.
+				if (!loop_header_block.complex_continue)
+					builder.add_block(loop_header_block.self);
+			}
 			else if (blocks.size() != 1 && is_single_block_loop(block))
 			{
 				// Awkward case, because the loop header is also the continue block.
@@ -3387,14 +3443,27 @@ void Compiler::analyze_variable_scope(SPIRFunction &entry, AnalyzeVariableScopeA
 
 			if (!first_use_is_dominator || force_temporary)
 			{
-				// This should be very rare, but if we try to declare a temporary inside a loop,
-				// and that temporary is used outside the loop as well (spirv-opt inliner likes this)
-				// we should actually emit the temporary outside the loop.
-				hoisted_temporaries.insert(var.first);
-				forced_temporaries.insert(var.first);
-
-				auto &block_temporaries = get<SPIRBlock>(dominating_block).declare_temporary;
-				block_temporaries.emplace_back(handler.result_id_to_type[var.first], var.first);
+				if (handler.access_chain_expressions.count(var.first))
+				{
+					// Exceptionally rare case.
+					// We cannot declare temporaries of access chains (except on MSL perhaps with pointers).
+					// Rather than do that, we force a complex loop to make sure access chains are created and consumed
+					// in expected order.
+					auto &loop_header_block = get<SPIRBlock>(dominating_block);
+					assert(loop_header_block.merge == SPIRBlock::MergeLoop);
+					loop_header_block.complex_continue = true;
+				}
+				else
+				{
+					// This should be very rare, but if we try to declare a temporary inside a loop,
+					// and that temporary is used outside the loop as well (spirv-opt inliner likes this)
+					// we should actually emit the temporary outside the loop.
+					hoisted_temporaries.insert(var.first);
+					forced_temporaries.insert(var.first);
+
+					auto &block_temporaries = get<SPIRBlock>(dominating_block).declare_temporary;
+					block_temporaries.emplace_back(handler.result_id_to_type[var.first], var.first);
+				}
 			}
 			else if (blocks.size() > 1)
 			{
@@ -3825,7 +3894,7 @@ void Compiler::build_function_control_flow_graphs_and_analyze()
 	}
 }
 
-Compiler::CFGBuilder::CFGBuilder(spirv_cross::Compiler &compiler_)
+Compiler::CFGBuilder::CFGBuilder(Compiler &compiler_)
     : compiler(compiler_)
 {
 }
@@ -3966,7 +4035,7 @@ void Compiler::make_constant_null(uint32_t id, uint32_t type)
 		if (!constant_type.array_size_literal.back())
 			SPIRV_CROSS_THROW("Array size of OpConstantNull must be a literal.");
 
-		vector<uint32_t> elements(constant_type.array.back());
+		SmallVector<uint32_t> elements(constant_type.array.back());
 		for (uint32_t i = 0; i < constant_type.array.back(); i++)
 			elements[i] = parent_id;
 		set<SPIRConstant>(id, type, elements.data(), uint32_t(elements.size()), false);
@@ -3974,7 +4043,7 @@ void Compiler::make_constant_null(uint32_t id, uint32_t type)
 	else if (!constant_type.member_types.empty())
 	{
 		uint32_t member_ids = ir.increase_bound_by(uint32_t(constant_type.member_types.size()));
-		vector<uint32_t> elements(constant_type.member_types.size());
+		SmallVector<uint32_t> elements(constant_type.member_types.size());
 		for (uint32_t i = 0; i < constant_type.member_types.size(); i++)
 		{
 			make_constant_null(member_ids + i, constant_type.member_types[i]);
@@ -3989,12 +4058,12 @@ void Compiler::make_constant_null(uint32_t id, uint32_t type)
 	}
 }
 
-const std::vector<spv::Capability> &Compiler::get_declared_capabilities() const
+const SmallVector<spv::Capability> &Compiler::get_declared_capabilities() const
 {
 	return ir.declared_capabilities;
 }
 
-const std::vector<std::string> &Compiler::get_declared_extensions() const
+const SmallVector<std::string> &Compiler::get_declared_extensions() const
 {
 	return ir.declared_extensions;
 }
@@ -4069,8 +4138,13 @@ Bitset Compiler::combined_decoration_for_member(const SPIRType &type, uint32_t i
 
 		// If our type is a struct, traverse all the members as well recursively.
 		flags.merge_or(dec.decoration_flags);
+
 		for (uint32_t i = 0; i < type.member_types.size(); i++)
-			flags.merge_or(combined_decoration_for_member(get<SPIRType>(type.member_types[i]), i));
+		{
+			auto &memb_type = get<SPIRType>(type.member_types[i]);
+			if (!memb_type.pointer)
+				flags.merge_or(combined_decoration_for_member(memb_type, i));
+		}
 	}
 
 	return flags;
@@ -4109,13 +4183,70 @@ bool Compiler::is_desktop_only_format(spv::ImageFormat format)
 	return false;
 }
 
-bool Compiler::image_is_comparison(const spirv_cross::SPIRType &type, uint32_t id) const
+bool Compiler::image_is_comparison(const SPIRType &type, uint32_t id) const
 {
 	return type.image.depth || (comparison_ids.count(id) != 0);
 }
 
-bool Compiler::type_is_opaque_value(const spirv_cross::SPIRType &type) const
+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;
+}
+
+Compiler::PhysicalStorageBufferPointerHandler::PhysicalStorageBufferPointerHandler(Compiler &compiler_)
+    : compiler(compiler_)
+{
+}
+
+bool Compiler::PhysicalStorageBufferPointerHandler::handle(Op op, const uint32_t *args, uint32_t)
+{
+	if (op == OpConvertUToPtr || op == OpBitcast)
+	{
+		auto &type = compiler.get<SPIRType>(args[0]);
+		if (type.storage == StorageClassPhysicalStorageBufferEXT && type.pointer && type.pointer_depth == 1)
+		{
+			// If we need to cast to a pointer type which is not a block, we might need to synthesize ourselves
+			// a block type which wraps this POD type.
+			if (type.basetype != SPIRType::Struct)
+				types.insert(args[0]);
+		}
+	}
+
+	return true;
+}
+
+void Compiler::analyze_non_block_pointer_types()
+{
+	PhysicalStorageBufferPointerHandler handler(*this);
+	traverse_all_reachable_opcodes(get<SPIRFunction>(ir.default_entry_point), handler);
+	physical_storage_non_block_pointer_types.reserve(handler.types.size());
+	for (auto type : handler.types)
+		physical_storage_non_block_pointer_types.push_back(type);
+	sort(begin(physical_storage_non_block_pointer_types), end(physical_storage_non_block_pointer_types));
+}
+
+bool Compiler::type_is_array_of_pointers(const SPIRType &type) const
+{
+	if (!type.pointer)
+		return false;
+
+	// If parent type has same pointer depth, we must have an array of pointers.
+	return type.pointer_depth == get<SPIRType>(type.parent_type).pointer_depth;
+}
diff --git a/src/3rdparty/SPIRV-Cross/spirv_cross.hpp b/src/3rdparty/SPIRV-Cross/spirv_cross.hpp
index 4edc836..4129e81 100644
--- a/src/3rdparty/SPIRV-Cross/spirv_cross.hpp
+++ b/src/3rdparty/SPIRV-Cross/spirv_cross.hpp
@@ -21,7 +21,7 @@
 #include "spirv_cfg.hpp"
 #include "spirv_cross_parsed_ir.hpp"
 
-namespace spirv_cross
+namespace SPIRV_CROSS_NAMESPACE
 {
 struct Resource
 {
@@ -54,24 +54,24 @@ struct Resource
 
 struct ShaderResources
 {
-	std::vector<Resource> uniform_buffers;
-	std::vector<Resource> storage_buffers;
-	std::vector<Resource> stage_inputs;
-	std::vector<Resource> stage_outputs;
-	std::vector<Resource> subpass_inputs;
-	std::vector<Resource> storage_images;
-	std::vector<Resource> sampled_images;
-	std::vector<Resource> atomic_counters;
-	std::vector<Resource> acceleration_structures;
+	SmallVector<Resource> uniform_buffers;
+	SmallVector<Resource> storage_buffers;
+	SmallVector<Resource> stage_inputs;
+	SmallVector<Resource> stage_outputs;
+	SmallVector<Resource> subpass_inputs;
+	SmallVector<Resource> storage_images;
+	SmallVector<Resource> sampled_images;
+	SmallVector<Resource> atomic_counters;
+	SmallVector<Resource> acceleration_structures;
 
 	// There can only be one push constant block,
 	// but keep the vector in case this restriction is lifted in the future.
-	std::vector<Resource> push_constant_buffers;
+	SmallVector<Resource> push_constant_buffers;
 
 	// For Vulkan GLSL and HLSL source,
 	// these correspond to separate texture2D and samplers respectively.
-	std::vector<Resource> separate_images;
-	std::vector<Resource> separate_samplers;
+	SmallVector<Resource> separate_images;
+	SmallVector<Resource> separate_samplers;
 };
 
 struct CombinedImageSampler
@@ -106,7 +106,9 @@ enum BufferPackingStandard
 	BufferPackingStd140EnhancedLayout,
 	BufferPackingStd430EnhancedLayout,
 	BufferPackingHLSLCbuffer,
-	BufferPackingHLSLCbufferPackOffset
+	BufferPackingHLSLCbufferPackOffset,
+	BufferPackingScalar,
+	BufferPackingScalarEnhancedLayout
 };
 
 struct EntryPoint
@@ -235,7 +237,7 @@ public:
 	// SPIR-V shader. The granularity of this analysis is per-member of a struct.
 	// This can be used for Buffer (UBO), BufferBlock/StorageBuffer (SSBO) and PushConstant blocks.
 	// ID is the Resource::id obtained from get_shader_resources().
-	std::vector<BufferRange> get_active_buffer_ranges(uint32_t id) const;
+	SmallVector<BufferRange> get_active_buffer_ranges(uint32_t id) const;
 
 	// Returns the effective size of a buffer block.
 	size_t get_declared_struct_size(const SPIRType &struct_type) const;
@@ -308,7 +310,7 @@ public:
 	// New variants of entry point query and reflection.
 	// Names for entry points in the SPIR-V module may alias if they belong to different execution models.
 	// To disambiguate, we must pass along with the entry point names the execution model.
-	std::vector<EntryPoint> get_entry_points_and_stages() const;
+	SmallVector<EntryPoint> get_entry_points_and_stages() const;
 	void set_entry_point(const std::string &entry, spv::ExecutionModel execution_model);
 
 	// Renames an entry point from old_name to new_name.
@@ -392,7 +394,7 @@ public:
 	void build_combined_image_samplers();
 
 	// Gets a remapping for the combined image samplers.
-	const std::vector<CombinedImageSampler> &get_combined_image_samplers() const
+	const SmallVector<CombinedImageSampler> &get_combined_image_samplers() const
 	{
 		return combined_image_samplers;
 	}
@@ -417,7 +419,7 @@ public:
 	// For composite types, the subconstants can be iterated over and modified.
 	// constant_type is the SPIRType for the specialization constant,
 	// which can be queried to determine which fields in the unions should be poked at.
-	std::vector<SpecializationConstant> get_specialization_constants() const;
+	SmallVector<SpecializationConstant> get_specialization_constants() const;
 	SPIRConstant &get_constant(uint32_t id);
 	const SPIRConstant &get_constant(uint32_t id) const;
 
@@ -468,10 +470,10 @@ public:
 	bool buffer_get_hlsl_counter_buffer(uint32_t id, uint32_t &counter_id) const;
 
 	// Gets the list of all SPIR-V Capabilities which were declared in the SPIR-V module.
-	const std::vector<spv::Capability> &get_declared_capabilities() const;
+	const SmallVector<spv::Capability> &get_declared_capabilities() const;
 
 	// Gets the list of all SPIR-V extensions which were declared in the SPIR-V module.
-	const std::vector<std::string> &get_declared_extensions() const;
+	const SmallVector<std::string> &get_declared_extensions() const;
 
 	// When declaring buffer blocks in GLSL, the name declared in the GLSL source
 	// might not be the same as the name declared in the SPIR-V module due to naming conflicts.
@@ -511,8 +513,8 @@ protected:
 	ParsedIR ir;
 	// Marks variables which have global scope and variables which can alias with other variables
 	// (SSBO, image load store, etc)
-	std::vector<uint32_t> global_variables;
-	std::vector<uint32_t> aliased_variables;
+	SmallVector<uint32_t> global_variables;
+	SmallVector<uint32_t> aliased_variables;
 
 	SPIRFunction *current_function = nullptr;
 	SPIRBlock *current_block = nullptr;
@@ -670,7 +672,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;
 
@@ -683,7 +688,7 @@ protected:
 	// variable is part of that entry points interface.
 	bool interface_variable_exists_in_entry_point(uint32_t id) const;
 
-	std::vector<CombinedImageSampler> combined_image_samplers;
+	SmallVector<CombinedImageSampler> combined_image_samplers;
 
 	void remap_variable_type_name(const SPIRType &type, const std::string &var_name, std::string &type_name) const
 	{
@@ -726,7 +731,7 @@ protected:
 
 	struct BufferAccessHandler : OpcodeHandler
 	{
-		BufferAccessHandler(const Compiler &compiler_, std::vector<BufferRange> &ranges_, uint32_t id_)
+		BufferAccessHandler(const Compiler &compiler_, SmallVector<BufferRange> &ranges_, uint32_t id_)
 		    : compiler(compiler_)
 		    , ranges(ranges_)
 		    , id(id_)
@@ -736,7 +741,7 @@ protected:
 		bool handle(spv::Op opcode, const uint32_t *args, uint32_t length) override;
 
 		const Compiler &compiler;
-		std::vector<BufferRange> &ranges;
+		SmallVector<BufferRange> &ranges;
 		uint32_t id;
 
 		std::unordered_set<uint32_t> seen;
@@ -807,7 +812,7 @@ protected:
 	bool traverse_all_reachable_opcodes(const SPIRBlock &block, OpcodeHandler &handler) const;
 	bool traverse_all_reachable_opcodes(const SPIRFunction &block, OpcodeHandler &handler) const;
 	// This must be an ordered data structure so we always pick the same type aliases.
-	std::vector<uint32_t> global_struct_cache;
+	SmallVector<uint32_t> global_struct_cache;
 
 	ShaderResources get_shader_resources(const std::unordered_set<uint32_t> *active_variables) const;
 
@@ -913,6 +918,7 @@ protected:
 		std::unordered_map<uint32_t, uint32_t> result_id_to_type;
 		std::unordered_map<uint32_t, std::unordered_set<uint32_t>> complete_write_variables_to_block;
 		std::unordered_map<uint32_t, std::unordered_set<uint32_t>> partial_write_variables_to_block;
+		std::unordered_set<uint32_t> access_chain_expressions;
 		const SPIRBlock *current_block = nullptr;
 	};
 
@@ -928,6 +934,16 @@ protected:
 		uint32_t write_count = 0;
 	};
 
+	struct PhysicalStorageBufferPointerHandler : OpcodeHandler
+	{
+		PhysicalStorageBufferPointerHandler(Compiler &compiler_);
+		bool handle(spv::Op op, const uint32_t *args, uint32_t length) override;
+		Compiler &compiler;
+		std::unordered_set<uint32_t> types;
+	};
+	void analyze_non_block_pointer_types();
+	SmallVector<uint32_t> physical_storage_non_block_pointer_types;
+
 	void analyze_variable_scope(SPIRFunction &function, AnalyzeVariableScopeAccessHandler &handler);
 	void find_function_local_luts(SPIRFunction &function, const AnalyzeVariableScopeAccessHandler &handler,
 	                              bool single_function);
@@ -955,6 +971,8 @@ protected:
 	bool has_extended_member_decoration(uint32_t type, uint32_t index, ExtendedDecorations decoration) const;
 	void unset_extended_member_decoration(uint32_t type, uint32_t index, ExtendedDecorations decoration);
 
+	bool type_is_array_of_pointers(const SPIRType &type) const;
+
 private:
 	// Used only to implement the old deprecated get_entry_point() interface.
 	const SPIREntryPoint &get_first_entry_point(const std::string &name) const;
@@ -964,6 +982,6 @@ private:
 	bool type_is_block_like(const SPIRType &type) const;
 	bool type_is_opaque_value(const SPIRType &type) const;
 };
-} // namespace spirv_cross
+} // namespace SPIRV_CROSS_NAMESPACE
 
 #endif
diff --git a/src/3rdparty/SPIRV-Cross/spirv_cross_c.cpp b/src/3rdparty/SPIRV-Cross/spirv_cross_c.cpp
index f41d216..d3352d9 100644
--- a/src/3rdparty/SPIRV-Cross/spirv_cross_c.cpp
+++ b/src/3rdparty/SPIRV-Cross/spirv_cross_c.cpp
@@ -34,9 +34,9 @@
 #include "spirv_reflect.hpp"
 #endif
 #include "spirv_parser.hpp"
-#include <string.h>
 #include <memory>
 #include <new>
+#include <string.h>
 
 // clang-format off
 
@@ -63,7 +63,7 @@
 #endif
 
 using namespace std;
-using namespace spirv_cross;
+using namespace SPIRV_CROSS_NAMESPACE;
 
 struct ScratchMemoryAllocation
 {
@@ -88,7 +88,7 @@ struct StringAllocation : ScratchMemoryAllocation
 template <typename T>
 struct TemporaryBuffer : ScratchMemoryAllocation
 {
-	std::vector<T> buffer;
+	SmallVector<T> buffer;
 };
 
 template <typename T, typename... Ts>
@@ -100,7 +100,7 @@ static inline std::unique_ptr<T> spvc_allocate(Ts &&... ts)
 struct spvc_context_s
 {
 	string last_error;
-	vector<unique_ptr<ScratchMemoryAllocation>> allocations;
+	SmallVector<unique_ptr<ScratchMemoryAllocation>> allocations;
 	const char *allocate_name(const std::string &name);
 
 	spvc_error_callback callback = nullptr;
@@ -173,20 +173,20 @@ struct spvc_constant_s : SPIRConstant
 struct spvc_resources_s : ScratchMemoryAllocation
 {
 	spvc_context context = nullptr;
-	std::vector<spvc_reflected_resource> uniform_buffers;
-	std::vector<spvc_reflected_resource> storage_buffers;
-	std::vector<spvc_reflected_resource> stage_inputs;
-	std::vector<spvc_reflected_resource> stage_outputs;
-	std::vector<spvc_reflected_resource> subpass_inputs;
-	std::vector<spvc_reflected_resource> storage_images;
-	std::vector<spvc_reflected_resource> sampled_images;
-	std::vector<spvc_reflected_resource> atomic_counters;
-	std::vector<spvc_reflected_resource> push_constant_buffers;
-	std::vector<spvc_reflected_resource> separate_images;
-	std::vector<spvc_reflected_resource> separate_samplers;
-	std::vector<spvc_reflected_resource> acceleration_structures;
-
-	bool copy_resources(std::vector<spvc_reflected_resource> &outputs, const std::vector<Resource> &inputs);
+	SmallVector<spvc_reflected_resource> uniform_buffers;
+	SmallVector<spvc_reflected_resource> storage_buffers;
+	SmallVector<spvc_reflected_resource> stage_inputs;
+	SmallVector<spvc_reflected_resource> stage_outputs;
+	SmallVector<spvc_reflected_resource> subpass_inputs;
+	SmallVector<spvc_reflected_resource> storage_images;
+	SmallVector<spvc_reflected_resource> sampled_images;
+	SmallVector<spvc_reflected_resource> atomic_counters;
+	SmallVector<spvc_reflected_resource> push_constant_buffers;
+	SmallVector<spvc_reflected_resource> separate_images;
+	SmallVector<spvc_reflected_resource> separate_samplers;
+	SmallVector<spvc_reflected_resource> acceleration_structures;
+
+	bool copy_resources(SmallVector<spvc_reflected_resource> &outputs, const SmallVector<Resource> &inputs);
 	bool copy_resources(const ShaderResources &resources);
 };
 
@@ -442,6 +442,9 @@ spvc_result spvc_compiler_options_set_uint(spvc_compiler_options options, spvc_c
 	case SPVC_COMPILER_OPTION_GLSL_EMIT_PUSH_CONSTANT_AS_UNIFORM_BUFFER:
 		options->glsl.emit_push_constant_as_uniform_buffer = value != 0;
 		break;
+	case SPVC_COMPILER_OPTION_GLSL_EMIT_UNIFORM_BUFFER_AS_PLAIN_UNIFORMS:
+		options->glsl.emit_uniform_buffer_as_plain_uniforms = value != 0;
+		break;
 #endif
 
 #if SPIRV_CROSS_C_API_HLSL
@@ -526,6 +529,10 @@ spvc_result spvc_compiler_options_set_uint(spvc_compiler_options options, spvc_c
 	case SPVC_COMPILER_OPTION_MSL_ARGUMENT_BUFFERS:
 		options->msl.argument_buffers = value != 0;
 		break;
+
+	case SPVC_COMPILER_OPTION_MSL_TEXTURE_BUFFER_NATIVE:
+		options->msl.texture_buffer_native = value != 0;
+		break;
 #endif
 
 	default:
@@ -634,7 +641,7 @@ spvc_result spvc_compiler_hlsl_set_root_constants_layout(spvc_compiler compiler,
 	}
 
 	auto &hlsl = *static_cast<CompilerHLSL *>(compiler->compiler.get());
-	std::vector<RootConstants> roots;
+	vector<RootConstants> roots;
 	roots.reserve(count);
 	for (size_t i = 0; i < count; i++)
 	{
@@ -980,8 +987,8 @@ spvc_result spvc_compiler_compile(spvc_compiler compiler, const char **source)
 	SPVC_END_SAFE_SCOPE(compiler->context, SPVC_ERROR_UNSUPPORTED_SPIRV)
 }
 
-bool spvc_resources_s::copy_resources(std::vector<spvc_reflected_resource> &outputs,
-                                      const std::vector<Resource> &inputs)
+bool spvc_resources_s::copy_resources(SmallVector<spvc_reflected_resource> &outputs,
+                                      const SmallVector<Resource> &inputs)
 {
 	for (auto &i : inputs)
 	{
@@ -1117,7 +1124,7 @@ spvc_result spvc_resources_get_resource_list_for_type(spvc_resources resources,
                                                       const spvc_reflected_resource **resource_list,
                                                       size_t *resource_size)
 {
-	const std::vector<spvc_reflected_resource> *list = nullptr;
+	const SmallVector<spvc_reflected_resource> *list = nullptr;
 	switch (type)
 	{
 	case SPVC_RESOURCE_TYPE_UNIFORM_BUFFER:
@@ -1275,7 +1282,7 @@ spvc_result spvc_compiler_get_entry_points(spvc_compiler compiler, const spvc_en
 	SPVC_BEGIN_SAFE_SCOPE
 	{
 		auto entries = compiler->compiler->get_entry_points_and_stages();
-		std::vector<spvc_entry_point> translated;
+		SmallVector<spvc_entry_point> translated;
 		translated.reserve(entries.size());
 
 		for (auto &entry : entries)
@@ -1406,7 +1413,7 @@ unsigned spvc_type_get_bit_width(spvc_type type)
 	return type->width;
 }
 
-unsigned spvc_type_get_vector_size(spvc_type type)
+unsigned spvc_type_get_SmallVector_size(spvc_type type)
 {
 	return type->vecsize;
 }
@@ -1566,7 +1573,7 @@ spvc_result spvc_compiler_get_combined_image_samplers(spvc_compiler compiler,
 	SPVC_BEGIN_SAFE_SCOPE
 	{
 		auto combined = compiler->compiler->get_combined_image_samplers();
-		std::vector<spvc_combined_image_sampler> translated;
+		SmallVector<spvc_combined_image_sampler> translated;
 		translated.reserve(combined.size());
 		for (auto &c : combined)
 		{
@@ -1591,7 +1598,7 @@ spvc_result spvc_compiler_get_specialization_constants(spvc_compiler compiler,
 	SPVC_BEGIN_SAFE_SCOPE
 	{
 		auto spec_constants = compiler->compiler->get_specialization_constants();
-		std::vector<spvc_specialization_constant> translated;
+		SmallVector<spvc_specialization_constant> translated;
 		translated.reserve(spec_constants.size());
 		for (auto &c : spec_constants)
 		{
@@ -1743,7 +1750,7 @@ spvc_result spvc_compiler_get_declared_extensions(spvc_compiler compiler, const
 	SPVC_BEGIN_SAFE_SCOPE
 	{
 		auto &exts = compiler->compiler->get_declared_extensions();
-		std::vector<const char *> duped;
+		SmallVector<const char *> duped;
 		duped.reserve(exts.size());
 		for (auto &ext : exts)
 			duped.push_back(compiler->context->allocate_name(ext));
diff --git a/src/3rdparty/SPIRV-Cross/spirv_cross_c.h b/src/3rdparty/SPIRV-Cross/spirv_cross_c.h
index 5491a2e..9e10d07 100644
--- a/src/3rdparty/SPIRV-Cross/spirv_cross_c.h
+++ b/src/3rdparty/SPIRV-Cross/spirv_cross_c.h
@@ -33,7 +33,7 @@ extern "C" {
 /* Bumped if ABI or API breaks backwards compatibility. */
 #define SPVC_C_API_VERSION_MAJOR 0
 /* Bumped if APIs or enumerations are added in a backwards compatible way. */
-#define SPVC_C_API_VERSION_MINOR 5
+#define SPVC_C_API_VERSION_MINOR 7
 /* Bumped if internal implementation details change. */
 #define SPVC_C_API_VERSION_PATCH 0
 
@@ -424,6 +424,10 @@ typedef enum spvc_compiler_option
 
 	SPVC_COMPILER_OPTION_GLSL_EMIT_PUSH_CONSTANT_AS_UNIFORM_BUFFER = 33 | SPVC_COMPILER_OPTION_GLSL_BIT,
 
+	SPVC_COMPILER_OPTION_MSL_TEXTURE_BUFFER_NATIVE = 34 | SPVC_COMPILER_OPTION_MSL_BIT,
+
+	SPVC_COMPILER_OPTION_GLSL_EMIT_UNIFORM_BUFFER_AS_PLAIN_UNIFORMS = 35 | SPVC_COMPILER_OPTION_GLSL_BIT,
+
 	SPVC_COMPILER_OPTION_INT_MAX = 0x7fffffff
 } spvc_compiler_option;
 
diff --git a/src/3rdparty/SPIRV-Cross/spirv_cross_containers.hpp b/src/3rdparty/SPIRV-Cross/spirv_cross_containers.hpp
new file mode 100644
index 0000000..393f461
--- /dev/null
+++ b/src/3rdparty/SPIRV-Cross/spirv_cross_containers.hpp
@@ -0,0 +1,715 @@
+/*
+ * Copyright 2019 Hans-Kristian Arntzen
+ *
+ * Licensed under the Apache License, Version 2.0 (the "License");
+ * you may not use this file except in compliance with the License.
+ * You may obtain a copy of the License at
+ *
+ *     http://www.apache.org/licenses/LICENSE-2.0
+ *
+ * Unless required by applicable law or agreed to in writing, software
+ * distributed under the License is distributed on an "AS IS" BASIS,
+ * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+ * See the License for the specific language governing permissions and
+ * limitations under the License.
+ */
+
+#ifndef SPIRV_CROSS_CONTAINERS_HPP
+#define SPIRV_CROSS_CONTAINERS_HPP
+
+#include "spirv_cross_error_handling.hpp"
+#include <algorithm>
+#include <functional>
+#include <iterator>
+#include <memory>
+#include <stack>
+#include <stdint.h>
+#include <stdlib.h>
+#include <string.h>
+#include <type_traits>
+#include <unordered_map>
+#include <unordered_set>
+#include <utility>
+#include <vector>
+
+#ifdef SPIRV_CROSS_NAMESPACE_OVERRIDE
+#define SPIRV_CROSS_NAMESPACE SPIRV_CROSS_NAMESPACE_OVERRIDE
+#else
+#define SPIRV_CROSS_NAMESPACE spirv_cross
+#endif
+
+namespace SPIRV_CROSS_NAMESPACE
+{
+#ifndef SPIRV_CROSS_FORCE_STL_TYPES
+// std::aligned_storage does not support size == 0, so roll our own.
+template <typename T, size_t N>
+class AlignedBuffer
+{
+public:
+	T *data()
+	{
+#if defined(_MSC_VER) && _MSC_VER < 1900
+		// MSVC 2013 workarounds, sigh ...
+		// Only use this workaround on MSVC 2013 due to some confusion around default initialized unions.
+		// Spec seems to suggest the memory will be zero-initialized, which is *not* what we want.
+		return reinterpret_cast<T *>(u.aligned_char);
+#else
+		return reinterpret_cast<T *>(aligned_char);
+#endif
+	}
+
+private:
+#if defined(_MSC_VER) && _MSC_VER < 1900
+	// MSVC 2013 workarounds, sigh ...
+	union {
+		char aligned_char[sizeof(T) * N];
+		double dummy_aligner;
+	} u;
+#else
+	alignas(T) char aligned_char[sizeof(T) * N];
+#endif
+};
+
+template <typename T>
+class AlignedBuffer<T, 0>
+{
+public:
+	T *data()
+	{
+		return nullptr;
+	}
+};
+
+// An immutable version of SmallVector which erases type information about storage.
+template <typename T>
+class VectorView
+{
+public:
+	T &operator[](size_t i)
+	{
+		return ptr[i];
+	}
+
+	const T &operator[](size_t i) const
+	{
+		return ptr[i];
+	}
+
+	bool empty() const
+	{
+		return buffer_size == 0;
+	}
+
+	size_t size() const
+	{
+		return buffer_size;
+	}
+
+	T *data()
+	{
+		return ptr;
+	}
+
+	const T *data() const
+	{
+		return ptr;
+	}
+
+	T *begin()
+	{
+		return ptr;
+	}
+
+	T *end()
+	{
+		return ptr + buffer_size;
+	}
+
+	const T *begin() const
+	{
+		return ptr;
+	}
+
+	const T *end() const
+	{
+		return ptr + buffer_size;
+	}
+
+	T &front()
+	{
+		return ptr[0];
+	}
+
+	const T &front() const
+	{
+		return ptr[0];
+	}
+
+	T &back()
+	{
+		return ptr[buffer_size - 1];
+	}
+
+	const T &back() const
+	{
+		return ptr[buffer_size - 1];
+	}
+
+	// Makes it easier to consume SmallVector.
+#if defined(_MSC_VER) && _MSC_VER < 1900
+	explicit operator std::vector<T>() const
+	{
+		// Another MSVC 2013 workaround. It does not understand lvalue/rvalue qualified operations.
+		return std::vector<T>(ptr, ptr + buffer_size);
+	}
+#else
+	// Makes it easier to consume SmallVector.
+	explicit operator std::vector<T>() const &
+	{
+		return std::vector<T>(ptr, ptr + buffer_size);
+	}
+
+	// If we are converting as an r-value, we can pilfer our elements.
+	explicit operator std::vector<T>() &&
+	{
+		return std::vector<T>(std::make_move_iterator(ptr), std::make_move_iterator(ptr + buffer_size));
+	}
+#endif
+
+	// Avoid sliced copies. Base class should only be read as a reference.
+	VectorView(const VectorView &) = delete;
+	void operator=(const VectorView &) = delete;
+
+protected:
+	VectorView() = default;
+	T *ptr = nullptr;
+	size_t buffer_size = 0;
+};
+
+// Simple vector which supports up to N elements inline, without malloc/free.
+// We use a lot of throwaway vectors all over the place which triggers allocations.
+// This class only implements the subset of std::vector we need in SPIRV-Cross.
+// It is *NOT* a drop-in replacement in general projects.
+template <typename T, size_t N = 8>
+class SmallVector : public VectorView<T>
+{
+public:
+	SmallVector()
+	{
+		this->ptr = stack_storage.data();
+		buffer_capacity = N;
+	}
+
+	SmallVector(const T *arg_list_begin, const T *arg_list_end)
+	    : SmallVector()
+	{
+		auto count = size_t(arg_list_end - arg_list_begin);
+		reserve(count);
+		for (size_t i = 0; i < count; i++, arg_list_begin++)
+			new (&this->ptr[i]) T(*arg_list_begin);
+		this->buffer_size = count;
+	}
+
+	SmallVector(SmallVector &&other) SPIRV_CROSS_NOEXCEPT : SmallVector()
+	{
+		*this = std::move(other);
+	}
+
+	SmallVector &operator=(SmallVector &&other) SPIRV_CROSS_NOEXCEPT
+	{
+		clear();
+		if (other.ptr != other.stack_storage.data())
+		{
+			// Pilfer allocated pointer.
+			if (this->ptr != stack_storage.data())
+				free(this->ptr);
+			this->ptr = other.ptr;
+			this->buffer_size = other.buffer_size;
+			buffer_capacity = other.buffer_capacity;
+			other.ptr = nullptr;
+			other.buffer_size = 0;
+			other.buffer_capacity = 0;
+		}
+		else
+		{
+			// Need to move the stack contents individually.
+			reserve(other.buffer_size);
+			for (size_t i = 0; i < other.buffer_size; i++)
+			{
+				new (&this->ptr[i]) T(std::move(other.ptr[i]));
+				other.ptr[i].~T();
+			}
+			this->buffer_size = other.buffer_size;
+			other.buffer_size = 0;
+		}
+		return *this;
+	}
+
+	SmallVector(const SmallVector &other)
+	    : SmallVector()
+	{
+		*this = other;
+	}
+
+	SmallVector &operator=(const SmallVector &other)
+	{
+		clear();
+		reserve(other.buffer_size);
+		for (size_t i = 0; i < other.buffer_size; i++)
+			new (&this->ptr[i]) T(other.ptr[i]);
+		this->buffer_size = other.buffer_size;
+		return *this;
+	}
+
+	explicit SmallVector(size_t count)
+	    : SmallVector()
+	{
+		resize(count);
+	}
+
+	~SmallVector()
+	{
+		clear();
+		if (this->ptr != stack_storage.data())
+			free(this->ptr);
+	}
+
+	void clear()
+	{
+		for (size_t i = 0; i < this->buffer_size; i++)
+			this->ptr[i].~T();
+		this->buffer_size = 0;
+	}
+
+	void push_back(const T &t)
+	{
+		reserve(this->buffer_size + 1);
+		new (&this->ptr[this->buffer_size]) T(t);
+		this->buffer_size++;
+	}
+
+	void push_back(T &&t)
+	{
+		reserve(this->buffer_size + 1);
+		new (&this->ptr[this->buffer_size]) T(std::move(t));
+		this->buffer_size++;
+	}
+
+	void pop_back()
+	{
+		// Work around false positive warning on GCC 8.3.
+		// Calling pop_back on empty vector is undefined.
+		if (!this->empty())
+			resize(this->buffer_size - 1);
+	}
+
+	template <typename... Ts>
+	void emplace_back(Ts &&... ts)
+	{
+		reserve(this->buffer_size + 1);
+		new (&this->ptr[this->buffer_size]) T(std::forward<Ts>(ts)...);
+		this->buffer_size++;
+	}
+
+	void reserve(size_t count)
+	{
+		if (count > buffer_capacity)
+		{
+			size_t target_capacity = buffer_capacity;
+			if (target_capacity == 0)
+				target_capacity = 1;
+			if (target_capacity < N)
+				target_capacity = N;
+
+			while (target_capacity < count)
+				target_capacity <<= 1u;
+
+			T *new_buffer =
+			    target_capacity > N ? static_cast<T *>(malloc(target_capacity * sizeof(T))) : stack_storage.data();
+
+			if (!new_buffer)
+				SPIRV_CROSS_THROW("Out of memory.");
+
+			// In case for some reason two allocations both come from same stack.
+			if (new_buffer != this->ptr)
+			{
+				// We don't deal with types which can throw in move constructor.
+				for (size_t i = 0; i < this->buffer_size; i++)
+				{
+					new (&new_buffer[i]) T(std::move(this->ptr[i]));
+					this->ptr[i].~T();
+				}
+			}
+
+			if (this->ptr != stack_storage.data())
+				free(this->ptr);
+			this->ptr = new_buffer;
+			buffer_capacity = target_capacity;
+		}
+	}
+
+	void insert(T *itr, const T *insert_begin, const T *insert_end)
+	{
+		auto count = size_t(insert_end - insert_begin);
+		if (itr == this->end())
+		{
+			reserve(this->buffer_size + count);
+			for (size_t i = 0; i < count; i++, insert_begin++)
+				new (&this->ptr[this->buffer_size + i]) T(*insert_begin);
+			this->buffer_size += count;
+		}
+		else
+		{
+			if (this->buffer_size + count > buffer_capacity)
+			{
+				auto target_capacity = this->buffer_size + count;
+				if (target_capacity == 0)
+					target_capacity = 1;
+				if (target_capacity < N)
+					target_capacity = N;
+
+				while (target_capacity < count)
+					target_capacity <<= 1u;
+
+				// Need to allocate new buffer. Move everything to a new buffer.
+				T *new_buffer =
+				    target_capacity > N ? static_cast<T *>(malloc(target_capacity * sizeof(T))) : stack_storage.data();
+				if (!new_buffer)
+					SPIRV_CROSS_THROW("Out of memory.");
+
+				// First, move elements from source buffer to new buffer.
+				// We don't deal with types which can throw in move constructor.
+				auto *target_itr = new_buffer;
+				auto *original_source_itr = this->begin();
+
+				if (new_buffer != this->ptr)
+				{
+					while (original_source_itr != itr)
+					{
+						new (target_itr) T(std::move(*original_source_itr));
+						original_source_itr->~T();
+						++original_source_itr;
+						++target_itr;
+					}
+				}
+
+				// Copy-construct new elements.
+				for (auto *source_itr = insert_begin; source_itr != insert_end; ++source_itr, ++target_itr)
+					new (target_itr) T(*source_itr);
+
+				// Move over the other half.
+				if (new_buffer != this->ptr || insert_begin != insert_end)
+				{
+					while (original_source_itr != this->end())
+					{
+						new (target_itr) T(std::move(*original_source_itr));
+						original_source_itr->~T();
+						++original_source_itr;
+						++target_itr;
+					}
+				}
+
+				if (this->ptr != stack_storage.data())
+					free(this->ptr);
+				this->ptr = new_buffer;
+				buffer_capacity = target_capacity;
+			}
+			else
+			{
+				// Move in place, need to be a bit careful about which elements are constructed and which are not.
+				// Move the end and construct the new elements.
+				auto *target_itr = this->end() + count;
+				auto *source_itr = this->end();
+				while (target_itr != this->end() && source_itr != itr)
+				{
+					--target_itr;
+					--source_itr;
+					new (target_itr) T(std::move(*source_itr));
+				}
+
+				// For already constructed elements we can move-assign.
+				std::move_backward(itr, source_itr, target_itr);
+
+				// For the inserts which go to already constructed elements, we can do a plain copy.
+				while (itr != this->end() && insert_begin != insert_end)
+					*itr++ = *insert_begin++;
+
+				// For inserts into newly allocated memory, we must copy-construct instead.
+				while (insert_begin != insert_end)
+				{
+					new (itr) T(*insert_begin);
+					++itr;
+					++insert_begin;
+				}
+			}
+
+			this->buffer_size += count;
+		}
+	}
+
+	T *erase(T *itr)
+	{
+		std::move(itr + 1, this->end(), itr);
+		this->ptr[--this->buffer_size].~T();
+		return itr;
+	}
+
+	void erase(T *start_erase, T *end_erase)
+	{
+		if (end_erase == this->end())
+		{
+			resize(size_t(start_erase - this->begin()));
+		}
+		else
+		{
+			auto new_size = this->buffer_size - (end_erase - start_erase);
+			std::move(end_erase, this->end(), start_erase);
+			resize(new_size);
+		}
+	}
+
+	void resize(size_t new_size)
+	{
+		if (new_size < this->buffer_size)
+		{
+			for (size_t i = new_size; i < this->buffer_size; i++)
+				this->ptr[i].~T();
+		}
+		else if (new_size > this->buffer_size)
+		{
+			reserve(new_size);
+			for (size_t i = this->buffer_size; i < new_size; i++)
+				new (&this->ptr[i]) T();
+		}
+
+		this->buffer_size = new_size;
+	}
+
+private:
+	size_t buffer_capacity = 0;
+	AlignedBuffer<T, N> stack_storage;
+};
+
+// A vector without stack storage.
+// Could also be a typedef-ed to std::vector,
+// but might as well use the one we have.
+template <typename T>
+using Vector = SmallVector<T, 0>;
+
+#else // SPIRV_CROSS_FORCE_STL_TYPES
+
+template <typename T, size_t N = 8>
+using SmallVector = std::vector<T>;
+template <typename T>
+using Vector = std::vector<T>;
+template <typename T>
+using VectorView = std::vector<T>;
+
+#endif // SPIRV_CROSS_FORCE_STL_TYPES
+
+// An object pool which we use for allocating IVariant-derived objects.
+// We know we are going to allocate a bunch of objects of each type,
+// so amortize the mallocs.
+class ObjectPoolBase
+{
+public:
+	virtual ~ObjectPoolBase() = default;
+	virtual void free_opaque(void *ptr) = 0;
+};
+
+template <typename T>
+class ObjectPool : public ObjectPoolBase
+{
+public:
+	explicit ObjectPool(unsigned start_object_count_ = 16)
+	    : start_object_count(start_object_count_)
+	{
+	}
+
+	template <typename... P>
+	T *allocate(P &&... p)
+	{
+		if (vacants.empty())
+		{
+			unsigned num_objects = start_object_count << memory.size();
+			T *ptr = static_cast<T *>(malloc(num_objects * sizeof(T)));
+			if (!ptr)
+				return nullptr;
+
+			for (unsigned i = 0; i < num_objects; i++)
+				vacants.push_back(&ptr[i]);
+
+			memory.emplace_back(ptr);
+		}
+
+		T *ptr = vacants.back();
+		vacants.pop_back();
+		new (ptr) T(std::forward<P>(p)...);
+		return ptr;
+	}
+
+	void free(T *ptr)
+	{
+		ptr->~T();
+		vacants.push_back(ptr);
+	}
+
+	void free_opaque(void *ptr) override
+	{
+		free(static_cast<T *>(ptr));
+	}
+
+	void clear()
+	{
+		vacants.clear();
+		memory.clear();
+	}
+
+protected:
+	Vector<T *> vacants;
+
+	struct MallocDeleter
+	{
+		void operator()(T *ptr)
+		{
+			::free(ptr);
+		}
+	};
+
+	SmallVector<std::unique_ptr<T, MallocDeleter>> memory;
+	unsigned start_object_count;
+};
+
+template <size_t StackSize = 4096, size_t BlockSize = 4096>
+class StringStream
+{
+public:
+	StringStream()
+	{
+		reset();
+	}
+
+	~StringStream()
+	{
+		reset();
+	}
+
+	// Disable copies and moves. Makes it easier to implement, and we don't need it.
+	StringStream(const StringStream &) = delete;
+	void operator=(const StringStream &) = delete;
+
+	template <typename T, typename std::enable_if<!std::is_floating_point<T>::value, int>::type = 0>
+	StringStream &operator<<(const T &t)
+	{
+		auto s = std::to_string(t);
+		append(s.data(), s.size());
+		return *this;
+	}
+
+	// Only overload this to make float/double conversions ambiguous.
+	StringStream &operator<<(uint32_t v)
+	{
+		auto s = std::to_string(v);
+		append(s.data(), s.size());
+		return *this;
+	}
+
+	StringStream &operator<<(char c)
+	{
+		append(&c, 1);
+		return *this;
+	}
+
+	StringStream &operator<<(const std::string &s)
+	{
+		append(s.data(), s.size());
+		return *this;
+	}
+
+	StringStream &operator<<(const char *s)
+	{
+		append(s, strlen(s));
+		return *this;
+	}
+
+	template <size_t N>
+	StringStream &operator<<(const char (&s)[N])
+	{
+		append(s, strlen(s));
+		return *this;
+	}
+
+	std::string str() const
+	{
+		std::string ret;
+		size_t target_size = 0;
+		for (auto &saved : saved_buffers)
+			target_size += saved.offset;
+		target_size += current_buffer.offset;
+		ret.reserve(target_size);
+
+		for (auto &saved : saved_buffers)
+			ret.insert(ret.end(), saved.buffer, saved.buffer + saved.offset);
+		ret.insert(ret.end(), current_buffer.buffer, current_buffer.buffer + current_buffer.offset);
+		return ret;
+	}
+
+	void reset()
+	{
+		for (auto &saved : saved_buffers)
+			if (saved.buffer != stack_buffer)
+				free(saved.buffer);
+		if (current_buffer.buffer != stack_buffer)
+			free(current_buffer.buffer);
+
+		saved_buffers.clear();
+		current_buffer.buffer = stack_buffer;
+		current_buffer.offset = 0;
+		current_buffer.size = sizeof(stack_buffer);
+	}
+
+private:
+	struct Buffer
+	{
+		char *buffer = nullptr;
+		size_t offset = 0;
+		size_t size = 0;
+	};
+	Buffer current_buffer;
+	char stack_buffer[StackSize];
+	SmallVector<Buffer> saved_buffers;
+
+	void append(const char *s, size_t len)
+	{
+		size_t avail = current_buffer.size - current_buffer.offset;
+		if (avail < len)
+		{
+			if (avail > 0)
+			{
+				memcpy(current_buffer.buffer + current_buffer.offset, s, avail);
+				s += avail;
+				len -= avail;
+				current_buffer.offset += avail;
+			}
+
+			saved_buffers.push_back(current_buffer);
+			size_t target_size = len > BlockSize ? len : BlockSize;
+			current_buffer.buffer = static_cast<char *>(malloc(target_size));
+			if (!current_buffer.buffer)
+				SPIRV_CROSS_THROW("Out of memory.");
+
+			memcpy(current_buffer.buffer, s, len);
+			current_buffer.offset = len;
+			current_buffer.size = target_size;
+		}
+		else
+		{
+			memcpy(current_buffer.buffer + current_buffer.offset, s, len);
+			current_buffer.offset += len;
+		}
+	}
+};
+
+} // namespace SPIRV_CROSS_NAMESPACE
+
+#endif
diff --git a/src/3rdparty/SPIRV-Cross/spirv_cross_error_handling.hpp b/src/3rdparty/SPIRV-Cross/spirv_cross_error_handling.hpp
new file mode 100644
index 0000000..e821c04
--- /dev/null
+++ b/src/3rdparty/SPIRV-Cross/spirv_cross_error_handling.hpp
@@ -0,0 +1,83 @@
+/*
+ * Copyright 2015-2019 Arm Limited
+ *
+ * Licensed under the Apache License, Version 2.0 (the "License");
+ * you may not use this file except in compliance with the License.
+ * You may obtain a copy of the License at
+ *
+ *     http://www.apache.org/licenses/LICENSE-2.0
+ *
+ * Unless required by applicable law or agreed to in writing, software
+ * distributed under the License is distributed on an "AS IS" BASIS,
+ * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+ * See the License for the specific language governing permissions and
+ * limitations under the License.
+ */
+
+#ifndef SPIRV_CROSS_ERROR_HANDLING
+#define SPIRV_CROSS_ERROR_HANDLING
+
+#include <stdexcept>
+#include <stdio.h>
+#include <stdlib.h>
+#include <string>
+
+#ifdef SPIRV_CROSS_NAMESPACE_OVERRIDE
+#define SPIRV_CROSS_NAMESPACE SPIRV_CROSS_NAMESPACE_OVERRIDE
+#else
+#define SPIRV_CROSS_NAMESPACE spirv_cross
+#endif
+
+namespace SPIRV_CROSS_NAMESPACE
+{
+#ifdef SPIRV_CROSS_EXCEPTIONS_TO_ASSERTIONS
+#if !defined(_MSC_VER) || defined(__clang__)
+[[noreturn]]
+#endif
+inline void
+report_and_abort(const std::string &msg)
+{
+#ifdef NDEBUG
+	(void)msg;
+#else
+	fprintf(stderr, "There was a compiler error: %s\n", msg.c_str());
+#endif
+	fflush(stderr);
+	abort();
+}
+
+#define SPIRV_CROSS_THROW(x) report_and_abort(x)
+#else
+class CompilerError : public std::runtime_error
+{
+public:
+	explicit CompilerError(const std::string &str)
+	    : std::runtime_error(str)
+	{
+	}
+};
+
+#define SPIRV_CROSS_THROW(x) throw CompilerError(x)
+#endif
+
+// MSVC 2013 does not have noexcept. We need this for Variant to get move constructor to work correctly
+// instead of copy constructor.
+// MSVC 2013 ignores that move constructors cannot throw in std::vector, so just don't define it.
+#if defined(_MSC_VER) && _MSC_VER < 1900
+#define SPIRV_CROSS_NOEXCEPT
+#else
+#define SPIRV_CROSS_NOEXCEPT noexcept
+#endif
+
+#if __cplusplus >= 201402l
+#define SPIRV_CROSS_DEPRECATED(reason) [[deprecated(reason)]]
+#elif defined(__GNUC__)
+#define SPIRV_CROSS_DEPRECATED(reason) __attribute__((deprecated))
+#elif defined(_MSC_VER)
+#define SPIRV_CROSS_DEPRECATED(reason) __declspec(deprecated(reason))
+#else
+#define SPIRV_CROSS_DEPRECATED(reason)
+#endif
+} // namespace SPIRV_CROSS_NAMESPACE
+
+#endif
diff --git a/src/3rdparty/SPIRV-Cross/spirv_cross_parsed_ir.cpp b/src/3rdparty/SPIRV-Cross/spirv_cross_parsed_ir.cpp
index f17c2be..108000c 100644
--- a/src/3rdparty/SPIRV-Cross/spirv_cross_parsed_ir.cpp
+++ b/src/3rdparty/SPIRV-Cross/spirv_cross_parsed_ir.cpp
@@ -21,11 +21,107 @@
 using namespace std;
 using namespace spv;
 
-namespace spirv_cross
+namespace SPIRV_CROSS_NAMESPACE
 {
+ParsedIR::ParsedIR()
+{
+	// If we move ParsedIR, we need to make sure the pointer stays fixed since the child Variant objects consume a pointer to this group,
+	// so need an extra pointer here.
+	pool_group.reset(new ObjectPoolGroup);
+
+	pool_group->pools[TypeType].reset(new ObjectPool<SPIRType>);
+	pool_group->pools[TypeVariable].reset(new ObjectPool<SPIRVariable>);
+	pool_group->pools[TypeConstant].reset(new ObjectPool<SPIRConstant>);
+	pool_group->pools[TypeFunction].reset(new ObjectPool<SPIRFunction>);
+	pool_group->pools[TypeFunctionPrototype].reset(new ObjectPool<SPIRFunctionPrototype>);
+	pool_group->pools[TypeBlock].reset(new ObjectPool<SPIRBlock>);
+	pool_group->pools[TypeExtension].reset(new ObjectPool<SPIRExtension>);
+	pool_group->pools[TypeExpression].reset(new ObjectPool<SPIRExpression>);
+	pool_group->pools[TypeConstantOp].reset(new ObjectPool<SPIRConstantOp>);
+	pool_group->pools[TypeCombinedImageSampler].reset(new ObjectPool<SPIRCombinedImageSampler>);
+	pool_group->pools[TypeAccessChain].reset(new ObjectPool<SPIRAccessChain>);
+	pool_group->pools[TypeUndef].reset(new ObjectPool<SPIRUndef>);
+}
+
+// Should have been default-implemented, but need this on MSVC 2013.
+ParsedIR::ParsedIR(ParsedIR &&other) SPIRV_CROSS_NOEXCEPT
+{
+	*this = move(other);
+}
+
+ParsedIR &ParsedIR::operator=(ParsedIR &&other) SPIRV_CROSS_NOEXCEPT
+{
+	if (this != &other)
+	{
+		pool_group = move(other.pool_group);
+		spirv = move(other.spirv);
+		meta = move(other.meta);
+		for (int i = 0; i < TypeCount; i++)
+			ids_for_type[i] = move(other.ids_for_type[i]);
+		ids_for_constant_or_type = move(other.ids_for_constant_or_type);
+		ids_for_constant_or_variable = move(other.ids_for_constant_or_variable);
+		declared_capabilities = move(other.declared_capabilities);
+		declared_extensions = move(other.declared_extensions);
+		block_meta = move(other.block_meta);
+		continue_block_to_loop_header = move(other.continue_block_to_loop_header);
+		entry_points = move(other.entry_points);
+		ids = move(other.ids);
+		addressing_model = other.addressing_model;
+		memory_model = other.memory_model;
+
+		default_entry_point = other.default_entry_point;
+		source = other.source;
+		loop_iteration_depth = other.loop_iteration_depth;
+	}
+	return *this;
+}
+
+ParsedIR::ParsedIR(const ParsedIR &other)
+    : ParsedIR()
+{
+	*this = other;
+}
+
+ParsedIR &ParsedIR::operator=(const ParsedIR &other)
+{
+	if (this != &other)
+	{
+		spirv = other.spirv;
+		meta = other.meta;
+		for (int i = 0; i < TypeCount; i++)
+			ids_for_type[i] = other.ids_for_type[i];
+		ids_for_constant_or_type = other.ids_for_constant_or_type;
+		ids_for_constant_or_variable = other.ids_for_constant_or_variable;
+		declared_capabilities = other.declared_capabilities;
+		declared_extensions = other.declared_extensions;
+		block_meta = other.block_meta;
+		continue_block_to_loop_header = other.continue_block_to_loop_header;
+		entry_points = other.entry_points;
+		default_entry_point = other.default_entry_point;
+		source = other.source;
+		loop_iteration_depth = other.loop_iteration_depth;
+		addressing_model = other.addressing_model;
+		memory_model = other.memory_model;
+
+		// Very deliberate copying of IDs. There is no default copy constructor, nor a simple default constructor.
+		// Construct object first so we have the correct allocator set-up, then we can copy object into our new pool group.
+		ids.clear();
+		ids.reserve(other.ids.size());
+		for (size_t i = 0; i < other.ids.size(); i++)
+		{
+			ids.emplace_back(pool_group.get());
+			ids.back() = other.ids[i];
+		}
+	}
+	return *this;
+}
+
 void ParsedIR::set_id_bounds(uint32_t bounds)
 {
-	ids.resize(bounds);
+	ids.reserve(bounds);
+	while (ids.size() < bounds)
+		ids.emplace_back(pool_group.get());
+
 	block_meta.resize(bounds);
 }
 
@@ -571,7 +667,11 @@ uint32_t ParsedIR::increase_bound_by(uint32_t incr_amount)
 {
 	auto curr_bound = ids.size();
 	auto new_bound = curr_bound + incr_amount;
-	ids.resize(new_bound);
+
+	ids.reserve(ids.size() + incr_amount);
+	for (uint32_t i = 0; i < incr_amount; i++)
+		ids.emplace_back(pool_group.get());
+
 	block_meta.resize(new_bound);
 	return uint32_t(curr_bound);
 }
@@ -596,24 +696,27 @@ void ParsedIR::add_typed_id(Types type, uint32_t id)
 	if (loop_iteration_depth)
 		SPIRV_CROSS_THROW("Cannot add typed ID while looping over it.");
 
-	switch (type)
+	if (ids[id].empty() || ids[id].get_type() != type)
 	{
-	case TypeConstant:
-		ids_for_constant_or_variable.push_back(id);
-		ids_for_constant_or_type.push_back(id);
-		break;
+		switch (type)
+		{
+		case TypeConstant:
+			ids_for_constant_or_variable.push_back(id);
+			ids_for_constant_or_type.push_back(id);
+			break;
 
-	case TypeVariable:
-		ids_for_constant_or_variable.push_back(id);
-		break;
+		case TypeVariable:
+			ids_for_constant_or_variable.push_back(id);
+			break;
 
-	case TypeType:
-	case TypeConstantOp:
-		ids_for_constant_or_type.push_back(id);
-		break;
+		case TypeType:
+		case TypeConstantOp:
+			ids_for_constant_or_type.push_back(id);
+			break;
 
-	default:
-		break;
+		default:
+			break;
+		}
 	}
 
 	if (ids[id].empty())
@@ -645,4 +748,4 @@ Meta *ParsedIR::find_meta(uint32_t id)
 		return nullptr;
 }
 
-} // namespace spirv_cross
+} // namespace SPIRV_CROSS_NAMESPACE
diff --git a/src/3rdparty/SPIRV-Cross/spirv_cross_parsed_ir.hpp b/src/3rdparty/SPIRV-Cross/spirv_cross_parsed_ir.hpp
index c3c4612..79e9e15 100644
--- a/src/3rdparty/SPIRV-Cross/spirv_cross_parsed_ir.hpp
+++ b/src/3rdparty/SPIRV-Cross/spirv_cross_parsed_ir.hpp
@@ -20,9 +20,8 @@
 #include "spirv_common.hpp"
 #include <stdint.h>
 #include <unordered_map>
-#include <vector>
 
-namespace spirv_cross
+namespace SPIRV_CROSS_NAMESPACE
 {
 
 // This data structure holds all information needed to perform cross-compilation and reflection.
@@ -32,7 +31,22 @@ namespace spirv_cross
 
 class ParsedIR
 {
+private:
+	// This must be destroyed after the "ids" vector.
+	std::unique_ptr<ObjectPoolGroup> pool_group;
+
 public:
+	ParsedIR();
+
+	// Due to custom allocations from object pools, we cannot use a default copy constructor.
+	ParsedIR(const ParsedIR &other);
+	ParsedIR &operator=(const ParsedIR &other);
+
+	// Moves are unproblematic, but we need to implement it anyways, since MSVC 2013 does not understand
+	// how to default-implement these.
+	ParsedIR(ParsedIR &&other) SPIRV_CROSS_NOEXCEPT;
+	ParsedIR &operator=(ParsedIR &&other) SPIRV_CROSS_NOEXCEPT;
+
 	// Resizes ids, meta and block_meta.
 	void set_id_bounds(uint32_t bounds);
 
@@ -40,7 +54,7 @@ public:
 	std::vector<uint32_t> spirv;
 
 	// Holds various data structures which inherit from IVariant.
-	std::vector<Variant> ids;
+	SmallVector<Variant> ids;
 
 	// Various meta data for IDs, decorations, names, etc.
 	std::unordered_map<uint32_t, Meta> meta;
@@ -48,19 +62,19 @@ public:
 	// Holds all IDs which have a certain type.
 	// This is needed so we can iterate through a specific kind of resource quickly,
 	// and in-order of module declaration.
-	std::vector<uint32_t> ids_for_type[TypeCount];
+	SmallVector<uint32_t> ids_for_type[TypeCount];
 
 	// Special purpose lists which contain a union of types.
 	// This is needed so we can declare specialization constants and structs in an interleaved fashion,
 	// among other things.
 	// Constants can be of struct type, and struct array sizes can use specialization constants.
-	std::vector<uint32_t> ids_for_constant_or_type;
-	std::vector<uint32_t> ids_for_constant_or_variable;
+	SmallVector<uint32_t> ids_for_constant_or_type;
+	SmallVector<uint32_t> ids_for_constant_or_variable;
 
 	// Declared capabilities and extensions in the SPIR-V module.
 	// Not really used except for reflection at the moment.
-	std::vector<spv::Capability> declared_capabilities;
-	std::vector<std::string> declared_extensions;
+	SmallVector<spv::Capability> declared_capabilities;
+	SmallVector<std::string> declared_extensions;
 
 	// Meta data about blocks. The cross-compiler needs to query if a block is either of these types.
 	// It is a bitset as there can be more than one tag per block.
@@ -73,7 +87,7 @@ public:
 		BLOCK_META_MULTISELECT_MERGE_BIT = 1 << 4
 	};
 	using BlockMetaFlags = uint8_t;
-	std::vector<BlockMetaFlags> block_meta;
+	SmallVector<BlockMetaFlags> block_meta;
 	std::unordered_map<uint32_t, uint32_t> continue_block_to_loop_header;
 
 	// Normally, we'd stick SPIREntryPoint in ids array, but it conflicts with SPIRFunction.
@@ -93,6 +107,9 @@ public:
 
 	Source source;
 
+	spv::AddressingModel addressing_model = spv::AddressingModelMax;
+	spv::MemoryModel memory_model = spv::MemoryModelMax;
+
 	// Decoration handling methods.
 	// Can be useful for simple "raw" reflection.
 	// However, most members are here because the Parser needs most of these,
@@ -181,6 +198,6 @@ private:
 	std::string empty_string;
 	Bitset cleared_bitset;
 };
-} // namespace spirv_cross
+} // namespace SPIRV_CROSS_NAMESPACE
 
 #endif
diff --git a/src/3rdparty/SPIRV-Cross/spirv_cross_util.cpp b/src/3rdparty/SPIRV-Cross/spirv_cross_util.cpp
index 58c1ddc..6ab5d26 100644
--- a/src/3rdparty/SPIRV-Cross/spirv_cross_util.cpp
+++ b/src/3rdparty/SPIRV-Cross/spirv_cross_util.cpp
@@ -18,12 +18,12 @@
 #include "spirv_common.hpp"
 
 using namespace spv;
-using namespace spirv_cross;
+using namespace SPIRV_CROSS_NAMESPACE;
 
 namespace spirv_cross_util
 {
-void rename_interface_variable(spirv_cross::Compiler &compiler, const std::vector<spirv_cross::Resource> &resources,
-                               uint32_t location, const std::string &name)
+void rename_interface_variable(Compiler &compiler, const SmallVector<Resource> &resources, uint32_t location,
+                               const std::string &name)
 {
 	for (auto &v : resources)
 	{
@@ -49,7 +49,7 @@ void rename_interface_variable(spirv_cross::Compiler &compiler, const std::vecto
 	}
 }
 
-void inherit_combined_sampler_bindings(spirv_cross::Compiler &compiler)
+void inherit_combined_sampler_bindings(Compiler &compiler)
 {
 	auto &samplers = compiler.get_combined_image_samplers();
 	for (auto &s : samplers)
diff --git a/src/3rdparty/SPIRV-Cross/spirv_cross_util.hpp b/src/3rdparty/SPIRV-Cross/spirv_cross_util.hpp
index faf0f48..7c4030b 100644
--- a/src/3rdparty/SPIRV-Cross/spirv_cross_util.hpp
+++ b/src/3rdparty/SPIRV-Cross/spirv_cross_util.hpp
@@ -21,9 +21,10 @@
 
 namespace spirv_cross_util
 {
-void rename_interface_variable(spirv_cross::Compiler &compiler, const std::vector<spirv_cross::Resource> &resources,
+void rename_interface_variable(SPIRV_CROSS_NAMESPACE::Compiler &compiler,
+                               const SPIRV_CROSS_NAMESPACE::SmallVector<SPIRV_CROSS_NAMESPACE::Resource> &resources,
                                uint32_t location, const std::string &name);
-void inherit_combined_sampler_bindings(spirv_cross::Compiler &compiler);
+void inherit_combined_sampler_bindings(SPIRV_CROSS_NAMESPACE::Compiler &compiler);
 } // namespace spirv_cross_util
 
 #endif
diff --git a/src/3rdparty/SPIRV-Cross/spirv_glsl.cpp b/src/3rdparty/SPIRV-Cross/spirv_glsl.cpp
index f35c7d8..32582fb 100644
--- a/src/3rdparty/SPIRV-Cross/spirv_glsl.cpp
+++ b/src/3rdparty/SPIRV-Cross/spirv_glsl.cpp
@@ -30,7 +30,7 @@
 #include <locale.h>
 
 using namespace spv;
-using namespace spirv_cross;
+using namespace SPIRV_CROSS_NAMESPACE;
 using namespace std;
 
 static bool is_unsigned_opcode(Op op)
@@ -106,6 +106,7 @@ static bool packing_has_flexible_offset(BufferPackingStandard packing)
 	{
 	case BufferPackingStd140:
 	case BufferPackingStd430:
+	case BufferPackingScalar:
 	case BufferPackingHLSLCbuffer:
 		return false;
 
@@ -114,6 +115,19 @@ static bool packing_has_flexible_offset(BufferPackingStandard packing)
 	}
 }
 
+static bool packing_is_scalar(BufferPackingStandard packing)
+{
+	switch (packing)
+	{
+	case BufferPackingScalar:
+	case BufferPackingScalarEnhancedLayout:
+		return true;
+
+	default:
+		return false;
+	}
+}
+
 static BufferPackingStandard packing_to_substruct_packing(BufferPackingStandard packing)
 {
 	switch (packing)
@@ -124,6 +138,8 @@ static BufferPackingStandard packing_to_substruct_packing(BufferPackingStandard
 		return BufferPackingStd430;
 	case BufferPackingHLSLCbufferPackOffset:
 		return BufferPackingHLSLCbuffer;
+	case BufferPackingScalarEnhancedLayout:
+		return BufferPackingScalar;
 	default:
 		return packing;
 	}
@@ -288,7 +304,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();
@@ -430,6 +446,44 @@ void CompilerGLSL::find_static_extensions()
 
 	if (options.separate_shader_objects && !options.es && options.version < 410)
 		require_extension_internal("GL_ARB_separate_shader_objects");
+
+	if (ir.addressing_model == AddressingModelPhysicalStorageBuffer64EXT)
+	{
+		if (!options.vulkan_semantics)
+			SPIRV_CROSS_THROW("GL_EXT_buffer_reference is only supported in Vulkan GLSL.");
+		if (options.es && options.version < 320)
+			SPIRV_CROSS_THROW("GL_EXT_buffer_reference requires ESSL 320.");
+		else if (!options.es && options.version < 450)
+			SPIRV_CROSS_THROW("GL_EXT_buffer_reference requires GLSL 450.");
+		require_extension_internal("GL_EXT_buffer_reference");
+	}
+	else if (ir.addressing_model != AddressingModelLogical)
+	{
+		SPIRV_CROSS_THROW("Only Logical and PhysicalStorageBuffer64EXT addressing models are supported.");
+	}
+
+	// Check for nonuniform qualifier.
+	// Instead of looping over all decorations to find this, just look at capabilities.
+	for (auto &cap : ir.declared_capabilities)
+	{
+		bool nonuniform_indexing = false;
+		switch (cap)
+		{
+		case CapabilityShaderNonUniformEXT:
+		case CapabilityRuntimeDescriptorArrayEXT:
+			if (!options.vulkan_semantics)
+				SPIRV_CROSS_THROW("GL_EXT_nonuniform_qualifier is only supported in Vulkan GLSL.");
+			require_extension_internal("GL_EXT_nonuniform_qualifier");
+			nonuniform_indexing = true;
+			break;
+
+		default:
+			break;
+		}
+
+		if (nonuniform_indexing)
+			break;
+	}
 }
 
 string CompilerGLSL::compile()
@@ -446,6 +500,11 @@ string CompilerGLSL::compile()
 	update_active_builtins();
 	analyze_image_and_sampler_usage();
 
+	// Shaders might cast unrelated data to pointers of non-block types.
+	// Find all such instances and make sure we can cast the pointers to a synthesized block type.
+	if (ir.addressing_model == AddressingModelPhysicalStorageBuffer64EXT)
+		analyze_non_block_pointer_types();
+
 	uint32_t pass_count = 0;
 	do
 	{
@@ -454,8 +513,7 @@ string CompilerGLSL::compile()
 
 		reset();
 
-		// Move constructor for this type is broken on GCC 4.9 ...
-		buffer = unique_ptr<ostringstream>(new ostringstream());
+		buffer.reset();
 
 		emit_header();
 		emit_resources();
@@ -463,20 +521,20 @@ 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";
 
-	return buffer->str();
+	return buffer.str();
 }
 
 std::string CompilerGLSL::get_partial_source()
 {
-	return buffer ? buffer->str() : "No compiled source available yet.";
+	return buffer.str();
 }
 
-void CompilerGLSL::build_workgroup_size(vector<string> &arguments, const SpecializationConstant &wg_x,
+void CompilerGLSL::build_workgroup_size(SmallVector<string> &arguments, const SpecializationConstant &wg_x,
                                         const SpecializationConstant &wg_y, const SpecializationConstant &wg_z)
 {
 	auto &execution = get_entry_point();
@@ -573,8 +631,8 @@ void CompilerGLSL::emit_header()
 	for (auto &header : header_lines)
 		statement(header);
 
-	vector<string> inputs;
-	vector<string> outputs;
+	SmallVector<string> inputs;
+	SmallVector<string> outputs;
 
 	switch (execution.model)
 	{
@@ -763,6 +821,8 @@ void CompilerGLSL::emit_struct(SPIRType &type)
 string CompilerGLSL::to_interpolation_qualifiers(const Bitset &flags)
 {
 	string res;
+	if (flags.get(DecorationNonUniformEXT))
+		res += "nonuniformEXT ";
 	//if (flags & (1ull << DecorationSmooth))
 	//    res += "smooth ";
 	if (flags.get(DecorationFlat))
@@ -798,7 +858,7 @@ string CompilerGLSL::layout_for_member(const SPIRType &type, uint32_t index)
 		return "";
 	auto &dec = memb[index];
 
-	vector<string> attr;
+	SmallVector<string> attr;
 
 	// We can only apply layouts on members in block interfaces.
 	// This is a bit problematic because in SPIR-V decorations are applied on the struct types directly.
@@ -973,6 +1033,24 @@ uint32_t CompilerGLSL::type_to_packed_base_size(const SPIRType &type, BufferPack
 uint32_t CompilerGLSL::type_to_packed_alignment(const SPIRType &type, const Bitset &flags,
                                                 BufferPackingStandard packing)
 {
+	// If using PhysicalStorageBufferEXT storage class, this is a pointer,
+	// and is 64-bit.
+	if (type.storage == StorageClassPhysicalStorageBufferEXT)
+	{
+		if (!type.pointer)
+			SPIRV_CROSS_THROW("Types in PhysicalStorageBufferEXT must be pointers.");
+
+		if (ir.addressing_model == AddressingModelPhysicalStorageBuffer64EXT)
+		{
+			if (packing_is_vec4_padded(packing) && type_is_array_of_pointers(type))
+				return 16;
+			else
+				return 8;
+		}
+		else
+			SPIRV_CROSS_THROW("AddressingModelPhysicalStorageBuffer64EXT must be used for PhysicalStorageBufferEXT.");
+	}
+
 	if (!type.array.empty())
 	{
 		uint32_t minimum_alignment = 1;
@@ -1008,6 +1086,10 @@ uint32_t CompilerGLSL::type_to_packed_alignment(const SPIRType &type, const Bits
 	{
 		const uint32_t base_alignment = type_to_packed_base_size(type, packing);
 
+		// Alignment requirement for scalar block layout is always the alignment for the most basic component.
+		if (packing_is_scalar(packing))
+			return base_alignment;
+
 		// Vectors are *not* aligned in HLSL, but there's an extra rule where vectors cannot straddle
 		// a vec4, this is handled outside since that part knows our current offset.
 		if (type.columns == 1 && packing_is_hlsl(packing))
@@ -1089,6 +1171,19 @@ uint32_t CompilerGLSL::type_to_packed_size(const SPIRType &type, const Bitset &f
 		return to_array_size_literal(type) * type_to_packed_array_stride(type, flags, packing);
 	}
 
+	// If using PhysicalStorageBufferEXT storage class, this is a pointer,
+	// and is 64-bit.
+	if (type.storage == StorageClassPhysicalStorageBufferEXT)
+	{
+		if (!type.pointer)
+			SPIRV_CROSS_THROW("Types in PhysicalStorageBufferEXT must be pointers.");
+
+		if (ir.addressing_model == AddressingModelPhysicalStorageBuffer64EXT)
+			return 8;
+		else
+			SPIRV_CROSS_THROW("AddressingModelPhysicalStorageBuffer64EXT must be used for PhysicalStorageBufferEXT.");
+	}
+
 	uint32_t size = 0;
 
 	if (type.basetype == SPIRType::Struct)
@@ -1118,27 +1213,34 @@ uint32_t CompilerGLSL::type_to_packed_size(const SPIRType &type, const Bitset &f
 	{
 		const uint32_t base_alignment = type_to_packed_base_size(type, packing);
 
-		if (type.columns == 1)
-			size = type.vecsize * base_alignment;
-
-		if (flags.get(DecorationColMajor) && type.columns > 1)
+		if (packing_is_scalar(packing))
 		{
-			if (packing_is_vec4_padded(packing))
-				size = type.columns * 4 * base_alignment;
-			else if (type.vecsize == 3)
-				size = type.columns * 4 * base_alignment;
-			else
-				size = type.columns * type.vecsize * base_alignment;
+			size = type.vecsize * type.columns * base_alignment;
 		}
-
-		if (flags.get(DecorationRowMajor) && type.vecsize > 1)
+		else
 		{
-			if (packing_is_vec4_padded(packing))
-				size = type.vecsize * 4 * base_alignment;
-			else if (type.columns == 3)
-				size = type.vecsize * 4 * base_alignment;
-			else
-				size = type.vecsize * type.columns * base_alignment;
+			if (type.columns == 1)
+				size = type.vecsize * base_alignment;
+
+			if (flags.get(DecorationColMajor) && type.columns > 1)
+			{
+				if (packing_is_vec4_padded(packing))
+					size = type.columns * 4 * base_alignment;
+				else if (type.vecsize == 3)
+					size = type.columns * 4 * base_alignment;
+				else
+					size = type.columns * type.vecsize * base_alignment;
+			}
+
+			if (flags.get(DecorationRowMajor) && type.vecsize > 1)
+			{
+				if (packing_is_vec4_padded(packing))
+					size = type.vecsize * 4 * base_alignment;
+				else if (type.columns == 3)
+					size = type.vecsize * 4 * base_alignment;
+				else
+					size = type.vecsize * type.columns * base_alignment;
+			}
 		}
 	}
 
@@ -1212,7 +1314,7 @@ bool CompilerGLSL::buffer_is_packing_standard(const SPIRType &type, BufferPackin
 
 		// The next member following a struct member is aligned to the base alignment of the struct that came before.
 		// GL 4.5 spec, 7.6.2.2.
-		if (memb_type.basetype == SPIRType::Struct)
+		if (memb_type.basetype == SPIRType::Struct && !memb_type.pointer)
 			pad_alignment = packed_alignment;
 		else
 			pad_alignment = 1;
@@ -1238,8 +1340,11 @@ bool CompilerGLSL::buffer_is_packing_standard(const SPIRType &type, BufferPackin
 			// We cannot use enhanced layouts on substructs, so they better be up to spec.
 			auto substruct_packing = packing_to_substruct_packing(packing);
 
-			if (!memb_type.member_types.empty() && !buffer_is_packing_standard(memb_type, substruct_packing))
+			if (!memb_type.pointer && !memb_type.member_types.empty() &&
+			    !buffer_is_packing_standard(memb_type, substruct_packing))
+			{
 				return false;
+			}
 		}
 
 		// Bump size.
@@ -1294,7 +1399,7 @@ string CompilerGLSL::layout_for_variable(const SPIRVariable &var)
 	if (is_legacy())
 		return "";
 
-	vector<string> attr;
+	SmallVector<string> attr;
 
 	auto &dec = ir.meta[var.self].decoration;
 	auto &type = get<SPIRType>(var.basetype);
@@ -1354,9 +1459,19 @@ string CompilerGLSL::layout_for_variable(const SPIRVariable &var)
 			attr.push_back(join("set = ", dec.set));
 	}
 
+	bool push_constant_block = options.vulkan_semantics && var.storage == StorageClassPushConstant;
+	bool ssbo_block = var.storage == StorageClassStorageBuffer ||
+	                  (var.storage == StorageClassUniform && typeflags.get(DecorationBufferBlock));
+	bool emulated_ubo = var.storage == StorageClassPushConstant && options.emit_push_constant_as_uniform_buffer;
+	bool ubo_block = var.storage == StorageClassUniform && typeflags.get(DecorationBlock);
+
 	// GL 3.0/GLSL 1.30 is not considered legacy, but it doesn't have UBOs ...
 	bool can_use_buffer_blocks = (options.es && options.version >= 300) || (!options.es && options.version >= 140);
 
+	// pretend no UBOs when options say so
+	if (ubo_block && options.emit_uniform_buffer_as_plain_uniforms)
+		can_use_buffer_blocks = false;
+
 	bool can_use_binding;
 	if (options.es)
 		can_use_binding = options.version >= 310;
@@ -1373,81 +1488,15 @@ string CompilerGLSL::layout_for_variable(const SPIRVariable &var)
 	if (flags.get(DecorationOffset))
 		attr.push_back(join("offset = ", dec.offset));
 
-	bool push_constant_block = options.vulkan_semantics && var.storage == StorageClassPushConstant;
-	bool ssbo_block = var.storage == StorageClassStorageBuffer ||
-	                  (var.storage == StorageClassUniform && typeflags.get(DecorationBufferBlock));
-	bool emulated_ubo = var.storage == StorageClassPushConstant && options.emit_push_constant_as_uniform_buffer;
-	bool ubo_block = var.storage == StorageClassUniform && typeflags.get(DecorationBlock);
-
 	// Instead of adding explicit offsets for every element here, just assume we're using std140 or std430.
 	// If SPIR-V does not comply with either layout, we cannot really work around it.
 	if (can_use_buffer_blocks && (ubo_block || emulated_ubo))
 	{
-		if (buffer_is_packing_standard(type, BufferPackingStd140))
-			attr.push_back("std140");
-		else if (buffer_is_packing_standard(type, BufferPackingStd140EnhancedLayout))
-		{
-			attr.push_back("std140");
-			// Fallback time. We might be able to use the ARB_enhanced_layouts to deal with this difference,
-			// however, we can only use layout(offset) on the block itself, not any substructs, so the substructs better be the appropriate layout.
-			// Enhanced layouts seem to always work in Vulkan GLSL, so no need for extensions there.
-			if (options.es && !options.vulkan_semantics)
-				SPIRV_CROSS_THROW("Uniform buffer block cannot be expressed as std140. ES-targets do "
-				                  "not support GL_ARB_enhanced_layouts.");
-			if (!options.es && !options.vulkan_semantics && options.version < 440)
-				require_extension_internal("GL_ARB_enhanced_layouts");
-
-			// This is a very last minute to check for this, but use this unused decoration to mark that we should emit
-			// explicit offsets for this block type.
-			// layout_for_variable() will be called before the actual buffer emit.
-			// The alternative is a full pass before codegen where we deduce this decoration,
-			// but then we are just doing the exact same work twice, and more complexity.
-			set_extended_decoration(type.self, SPIRVCrossDecorationPacked);
-		}
-		else
-		{
-			SPIRV_CROSS_THROW("Uniform buffer cannot be expressed as std140, even with enhanced layouts. You can try "
-			                  "flattening this block to "
-			                  "support a more flexible layout.");
-		}
+		attr.push_back(buffer_to_packing_standard(type, false));
 	}
 	else if (can_use_buffer_blocks && (push_constant_block || ssbo_block))
 	{
-		if (buffer_is_packing_standard(type, BufferPackingStd430))
-			attr.push_back("std430");
-		else if (buffer_is_packing_standard(type, BufferPackingStd140))
-			attr.push_back("std140");
-		else if (buffer_is_packing_standard(type, BufferPackingStd140EnhancedLayout))
-		{
-			attr.push_back("std140");
-
-			// Fallback time. We might be able to use the ARB_enhanced_layouts to deal with this difference,
-			// however, we can only use layout(offset) on the block itself, not any substructs, so the substructs better be the appropriate layout.
-			// Enhanced layouts seem to always work in Vulkan GLSL, so no need for extensions there.
-			if (options.es && !options.vulkan_semantics)
-				SPIRV_CROSS_THROW("Push constant block cannot be expressed as neither std430 nor std140. ES-targets do "
-				                  "not support GL_ARB_enhanced_layouts.");
-			if (!options.es && !options.vulkan_semantics && options.version < 440)
-				require_extension_internal("GL_ARB_enhanced_layouts");
-
-			set_extended_decoration(type.self, SPIRVCrossDecorationPacked);
-		}
-		else if (buffer_is_packing_standard(type, BufferPackingStd430EnhancedLayout))
-		{
-			attr.push_back("std430");
-			if (options.es && !options.vulkan_semantics)
-				SPIRV_CROSS_THROW("Push constant block cannot be expressed as neither std430 nor std140. ES-targets do "
-				                  "not support GL_ARB_enhanced_layouts.");
-			if (!options.es && !options.vulkan_semantics && options.version < 440)
-				require_extension_internal("GL_ARB_enhanced_layouts");
-
-			set_extended_decoration(type.self, SPIRVCrossDecorationPacked);
-		}
-		else
-		{
-			SPIRV_CROSS_THROW("Buffer block cannot be expressed as neither std430 nor std140, even with enhanced "
-			                  "layouts. You can try flattening this block to support a more flexible layout.");
-		}
+		attr.push_back(buffer_to_packing_standard(type, true));
 	}
 
 	// For images, the type itself adds a layout qualifer.
@@ -1468,6 +1517,55 @@ string CompilerGLSL::layout_for_variable(const SPIRVariable &var)
 	return res;
 }
 
+string CompilerGLSL::buffer_to_packing_standard(const SPIRType &type, bool check_std430)
+{
+	if (check_std430 && buffer_is_packing_standard(type, BufferPackingStd430))
+		return "std430";
+	else if (buffer_is_packing_standard(type, BufferPackingStd140))
+		return "std140";
+	else if (options.vulkan_semantics && buffer_is_packing_standard(type, BufferPackingScalar))
+	{
+		require_extension_internal("GL_EXT_scalar_block_layout");
+		return "scalar";
+	}
+	else if (check_std430 && buffer_is_packing_standard(type, BufferPackingStd430EnhancedLayout))
+	{
+		if (options.es && !options.vulkan_semantics)
+			SPIRV_CROSS_THROW("Push constant block cannot be expressed as neither std430 nor std140. ES-targets do "
+			                  "not support GL_ARB_enhanced_layouts.");
+		if (!options.es && !options.vulkan_semantics && options.version < 440)
+			require_extension_internal("GL_ARB_enhanced_layouts");
+
+		set_extended_decoration(type.self, SPIRVCrossDecorationPacked);
+		return "std430";
+	}
+	else if (buffer_is_packing_standard(type, BufferPackingStd140EnhancedLayout))
+	{
+		// Fallback time. We might be able to use the ARB_enhanced_layouts to deal with this difference,
+		// however, we can only use layout(offset) on the block itself, not any substructs, so the substructs better be the appropriate layout.
+		// Enhanced layouts seem to always work in Vulkan GLSL, so no need for extensions there.
+		if (options.es && !options.vulkan_semantics)
+			SPIRV_CROSS_THROW("Push constant block cannot be expressed as neither std430 nor std140. ES-targets do "
+			                  "not support GL_ARB_enhanced_layouts.");
+		if (!options.es && !options.vulkan_semantics && options.version < 440)
+			require_extension_internal("GL_ARB_enhanced_layouts");
+
+		set_extended_decoration(type.self, SPIRVCrossDecorationPacked);
+		return "std140";
+	}
+	else if (options.vulkan_semantics && buffer_is_packing_standard(type, BufferPackingScalarEnhancedLayout))
+	{
+		set_extended_decoration(type.self, SPIRVCrossDecorationPacked);
+		require_extension_internal("GL_EXT_scalar_block_layout");
+		return "scalar";
+	}
+	else
+	{
+		SPIRV_CROSS_THROW("Buffer block cannot be expressed as any of std430, std140, scalar, even with enhanced "
+		                  "layouts. You can try flattening this block to support a more flexible layout.");
+	}
+}
+
 void CompilerGLSL::emit_push_constant_block(const SPIRVariable &var)
 {
 	if (flattened_buffer_blocks.count(var.self))
@@ -1517,9 +1615,13 @@ void CompilerGLSL::emit_push_constant_block_glsl(const SPIRVariable &var)
 
 void CompilerGLSL::emit_buffer_block(const SPIRVariable &var)
 {
+	auto &type = get<SPIRType>(var.basetype);
+	bool ubo_block = var.storage == StorageClassUniform && has_decoration(type.self, DecorationBlock);
+
 	if (flattened_buffer_blocks.count(var.self))
 		emit_buffer_block_flattened(var);
-	else if (is_legacy() || (!options.es && options.version == 130))
+	else if (is_legacy() || (!options.es && options.version == 130) ||
+	         (ubo_block && options.emit_uniform_buffer_as_plain_uniforms))
 		emit_buffer_block_legacy(var);
 	else
 		emit_buffer_block_native(var);
@@ -1545,6 +1647,81 @@ void CompilerGLSL::emit_buffer_block_legacy(const SPIRVariable &var)
 	statement("");
 }
 
+void CompilerGLSL::emit_buffer_reference_block(SPIRType &type, bool forward_declaration)
+{
+	string buffer_name;
+
+	if (forward_declaration)
+	{
+		// Block names should never alias, but from HLSL input they kind of can because block types are reused for UAVs ...
+		// Allow aliased name since we might be declaring the block twice. Once with buffer reference (forward declared) and one proper declaration.
+		// The names must match up.
+		buffer_name = to_name(type.self, false);
+
+		// Shaders never use the block by interface name, so we don't
+		// have to track this other than updating name caches.
+		// If we have a collision for any reason, just fallback immediately.
+		if (ir.meta[type.self].decoration.alias.empty() ||
+		    block_ssbo_names.find(buffer_name) != end(block_ssbo_names) ||
+		    resource_names.find(buffer_name) != end(resource_names))
+		{
+			buffer_name = join("_", type.self);
+		}
+
+		// Make sure we get something unique for both global name scope and block name scope.
+		// See GLSL 4.5 spec: section 4.3.9 for details.
+		add_variable(block_ssbo_names, resource_names, buffer_name);
+
+		// If for some reason buffer_name is an illegal name, make a final fallback to a workaround name.
+		// This cannot conflict with anything else, so we're safe now.
+		// We cannot reuse this fallback name in neither global scope (blocked by block_names) nor block name scope.
+		if (buffer_name.empty())
+			buffer_name = join("_", type.self);
+
+		block_names.insert(buffer_name);
+		block_ssbo_names.insert(buffer_name);
+	}
+	else if (type.basetype != SPIRType::Struct)
+		buffer_name = type_to_glsl(type);
+	else
+		buffer_name = to_name(type.self, false);
+
+	if (!forward_declaration)
+	{
+		if (type.basetype == SPIRType::Struct)
+			statement("layout(buffer_reference, ", buffer_to_packing_standard(type, true), ") buffer ", buffer_name);
+		else
+			statement("layout(buffer_reference) buffer ", buffer_name);
+
+		begin_scope();
+
+		if (type.basetype == SPIRType::Struct)
+		{
+			type.member_name_cache.clear();
+
+			uint32_t i = 0;
+			for (auto &member : type.member_types)
+			{
+				add_member_name(type, i);
+				emit_struct_member(type, member, i);
+				i++;
+			}
+		}
+		else
+		{
+			auto &pointee_type = get_pointee_type(type);
+			statement(type_to_glsl(pointee_type), " value", type_to_array_glsl(pointee_type), ";");
+		}
+
+		end_scope_decl();
+		statement("");
+	}
+	else
+	{
+		statement("layout(buffer_reference) buffer ", buffer_name, ";");
+	}
+}
+
 void CompilerGLSL::emit_buffer_block_native(const SPIRVariable &var)
 {
 	auto &type = get<SPIRType>(var.basetype);
@@ -1630,7 +1807,7 @@ void CompilerGLSL::emit_buffer_block_flattened(const SPIRVariable &var)
 			SPIRV_CROSS_THROW("Basic types in a flattened UBO must be float, int or uint.");
 
 		auto flags = ir.get_buffer_block_flags(var);
-		statement("uniform ", flags_to_precision_qualifiers_glsl(tmp, flags), type_to_glsl(tmp), " ", buffer_name, "[",
+		statement("uniform ", flags_to_qualifiers_glsl(tmp, flags), type_to_glsl(tmp), " ", buffer_name, "[",
 		          buffer_size, "];");
 	}
 	else
@@ -2325,7 +2502,7 @@ void CompilerGLSL::emit_resources()
 
 		if ((wg_x.id != 0) || (wg_y.id != 0) || (wg_z.id != 0))
 		{
-			vector<string> inputs;
+			SmallVector<string> inputs;
 			build_workgroup_size(inputs, wg_x, wg_y, wg_z);
 			statement("layout(", merge(inputs), ") in;");
 			statement("");
@@ -2334,6 +2511,36 @@ void CompilerGLSL::emit_resources()
 
 	emitted = false;
 
+	if (ir.addressing_model == AddressingModelPhysicalStorageBuffer64EXT)
+	{
+		for (auto type : physical_storage_non_block_pointer_types)
+		{
+			emit_buffer_reference_block(get<SPIRType>(type), false);
+		}
+
+		// Output buffer reference blocks.
+		// Do this in two stages, one with forward declaration,
+		// and one without. Buffer reference blocks can reference themselves
+		// to support things like linked lists.
+		ir.for_each_typed_id<SPIRType>([&](uint32_t, SPIRType &type) {
+			bool has_block_flags = has_decoration(type.self, DecorationBlock);
+			if (has_block_flags && type.pointer && type.pointer_depth == 1 && !type_is_array_of_pointers(type) &&
+			    type.storage == StorageClassPhysicalStorageBufferEXT)
+			{
+				emit_buffer_reference_block(type, true);
+			}
+		});
+
+		ir.for_each_typed_id<SPIRType>([&](uint32_t, SPIRType &type) {
+			bool has_block_flags = has_decoration(type.self, DecorationBlock);
+			if (has_block_flags && type.pointer && type.pointer_depth == 1 && !type_is_array_of_pointers(type) &&
+			    type.storage == StorageClassPhysicalStorageBufferEXT)
+			{
+				emit_buffer_reference_block(type, false);
+			}
+		});
+	}
+
 	// Output UBOs and SSBOs
 	ir.for_each_typed_id<SPIRVariable>([&](uint32_t, SPIRVariable &var) {
 		auto &type = this->get<SPIRType>(var.basetype);
@@ -2454,7 +2661,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,
@@ -2535,15 +2742,22 @@ string CompilerGLSL::enclose_expression(const string &expr)
 		return expr;
 }
 
-string CompilerGLSL::dereference_expression(const std::string &expr)
+string CompilerGLSL::dereference_expression(const SPIRType &expr_type, const std::string &expr)
 {
 	// If this expression starts with an address-of operator ('&'), then
 	// just return the part after the operator.
 	// TODO: Strip parens if unnecessary?
 	if (expr.front() == '&')
 		return expr.substr(1);
-	else
+	else if (backend.native_pointers)
 		return join('*', expr);
+	else if (expr_type.storage == StorageClassPhysicalStorageBufferEXT && expr_type.basetype != SPIRType::Struct &&
+	         expr_type.pointer_depth == 1)
+	{
+		return join(enclose_expression(expr), ".value");
+	}
+	else
+		return expr;
 }
 
 string CompilerGLSL::address_of_expression(const std::string &expr)
@@ -2591,7 +2805,7 @@ string CompilerGLSL::to_dereferenced_expression(uint32_t id, bool register_expre
 {
 	auto &type = expression_type(id);
 	if (type.pointer && should_dereference(id))
-		return dereference_expression(to_enclosed_expression(id, register_expression_read));
+		return dereference_expression(type, to_enclosed_expression(id, register_expression_read));
 	else
 		return to_expression(id, register_expression_read);
 }
@@ -2665,7 +2879,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.
@@ -2869,7 +3083,7 @@ string CompilerGLSL::constant_op_expression(const SPIRConstantOp &cop)
 	}
 
 	uint32_t bit_width = 0;
-	if (unary || binary)
+	if (unary || binary || cop.opcode == OpSConvert || cop.opcode == OpUConvert)
 		bit_width = expression_type(cop.arguments[0]).width;
 
 	SPIRType::BaseType input_type;
@@ -2889,6 +3103,8 @@ string CompilerGLSL::constant_op_expression(const SPIRConstantOp &cop)
 	case OpSMod:
 	case OpSDiv:
 	case OpShiftRightArithmetic:
+	case OpSConvert:
+	case OpSNegate:
 		input_type = to_signed_basetype(bit_width);
 		break;
 
@@ -2899,6 +3115,7 @@ string CompilerGLSL::constant_op_expression(const SPIRConstantOp &cop)
 	case OpUMod:
 	case OpUDiv:
 	case OpShiftRightLogical:
+	case OpUConvert:
 		input_type = to_unsigned_basetype(bit_width);
 		break;
 
@@ -2940,6 +3157,21 @@ string CompilerGLSL::constant_op_expression(const SPIRConstantOp &cop)
 		// Works around various casting scenarios in glslang as there is no OpBitcast for specialization constants.
 		return join("(", op, bitcast_glsl(type, cop.arguments[0]), ")");
 	}
+	else if (cop.opcode == OpSConvert || cop.opcode == OpUConvert)
+	{
+		if (cop.arguments.size() < 1)
+			SPIRV_CROSS_THROW("Not enough arguments to OpSpecConstantOp.");
+
+		auto &arg_type = expression_type(cop.arguments[0]);
+		if (arg_type.width < type.width && input_type != arg_type.basetype)
+		{
+			auto expected = arg_type;
+			expected.basetype = input_type;
+			return join(op, "(", bitcast_glsl(expected, cop.arguments[0]), ")");
+		}
+		else
+			return join(op, "(", to_expression(cop.arguments[0]), ")");
+	}
 	else
 	{
 		if (cop.arguments.size() < 1)
@@ -3581,6 +3813,41 @@ string CompilerGLSL::constant_expression_vector(const SPIRConstant &c, uint32_t
 	return res;
 }
 
+SPIRExpression &CompilerGLSL::emit_uninitialized_temporary_expression(uint32_t type, uint32_t id)
+{
+	forced_temporaries.insert(id);
+	emit_uninitialized_temporary(type, id);
+	return set<SPIRExpression>(id, to_name(id), type, true);
+}
+
+void CompilerGLSL::emit_uninitialized_temporary(uint32_t result_type, uint32_t result_id)
+{
+	// If we're declaring temporaries inside continue blocks,
+	// we must declare the temporary in the loop header so that the continue block can avoid declaring new variables.
+	if (current_continue_block && !hoisted_temporaries.count(result_id))
+	{
+		auto &header = get<SPIRBlock>(current_continue_block->loop_dominator);
+		if (find_if(begin(header.declare_temporary), end(header.declare_temporary),
+		            [result_type, result_id](const pair<uint32_t, uint32_t> &tmp) {
+			            return tmp.first == result_type && tmp.second == result_id;
+		            }) == end(header.declare_temporary))
+		{
+			header.declare_temporary.emplace_back(result_type, result_id);
+			hoisted_temporaries.insert(result_id);
+			force_recompile();
+		}
+	}
+	else if (hoisted_temporaries.count(result_id) == 0)
+	{
+		auto &type = get<SPIRType>(result_type);
+		auto &flags = ir.meta[result_id].decoration.decoration_flags;
+
+		// The result_id has not been made into an expression yet, so use flags interface.
+		add_local_variable_name(result_id);
+		statement(flags_to_qualifiers_glsl(type, flags), variable_decl(type, to_name(result_id)), ";");
+	}
+}
+
 string CompilerGLSL::declare_temporary(uint32_t result_type, uint32_t result_id)
 {
 	auto &type = get<SPIRType>(result_type);
@@ -3598,7 +3865,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), " = ");
@@ -3612,7 +3879,7 @@ string CompilerGLSL::declare_temporary(uint32_t result_type, uint32_t result_id)
 	{
 		// The result_id has not been made into an expression yet, so use flags interface.
 		add_local_variable_name(result_id);
-		return join(flags_to_precision_qualifiers_glsl(type, flags), variable_decl(type, to_name(result_id)), " = ");
+		return join(flags_to_qualifiers_glsl(type, flags), variable_decl(type, to_name(result_id)), " = ");
 	}
 }
 
@@ -3788,14 +4055,19 @@ void CompilerGLSL::emit_unary_func_op_cast(uint32_t result_type, uint32_t result
                                            SPIRType::BaseType input_type, SPIRType::BaseType expected_result_type)
 {
 	auto &out_type = get<SPIRType>(result_type);
+	auto &expr_type = expression_type(op0);
 	auto expected_type = out_type;
+
+	// Bit-widths might be different in unary cases because we use it for SConvert/UConvert and friends.
 	expected_type.basetype = input_type;
-	string cast_op = expression_type(op0).basetype != input_type ? bitcast_glsl(expected_type, op0) : to_expression(op0);
+	expected_type.width = expr_type.width;
+	string cast_op = expr_type.basetype != input_type ? bitcast_glsl(expected_type, op0) : to_unpacked_expression(op0);
 
 	string expr;
 	if (out_type.basetype != expected_result_type)
 	{
 		expected_type.basetype = expected_result_type;
+		expected_type.width = out_type.width;
 		expr = bitcast_glsl_op(out_type, expected_type);
 		expr += '(';
 		expr += join(op, "(", cast_op, ")");
@@ -3810,17 +4082,18 @@ void CompilerGLSL::emit_unary_func_op_cast(uint32_t result_type, uint32_t result
 	inherit_expression_dependencies(result_id, op0);
 }
 
-void CompilerGLSL::emit_trinary_func_op_cast(uint32_t result_type, uint32_t result_id,
-                                             uint32_t op0, uint32_t op1, uint32_t op2,
-                                             const char *op,
-                                             SPIRType::BaseType input_type)
+void CompilerGLSL::emit_trinary_func_op_cast(uint32_t result_type, uint32_t result_id, uint32_t op0, uint32_t op1,
+                                             uint32_t op2, const char *op, SPIRType::BaseType input_type)
 {
 	auto &out_type = get<SPIRType>(result_type);
 	auto expected_type = out_type;
 	expected_type.basetype = input_type;
-	string cast_op0 = expression_type(op0).basetype != input_type ? bitcast_glsl(expected_type, op0) : to_expression(op0);
-	string cast_op1 = expression_type(op1).basetype != input_type ? bitcast_glsl(expected_type, op1) : to_expression(op1);
-	string cast_op2 = expression_type(op2).basetype != input_type ? bitcast_glsl(expected_type, op2) : to_expression(op2);
+	string cast_op0 =
+	    expression_type(op0).basetype != input_type ? bitcast_glsl(expected_type, op0) : to_unpacked_expression(op0);
+	string cast_op1 =
+	    expression_type(op1).basetype != input_type ? bitcast_glsl(expected_type, op1) : to_unpacked_expression(op1);
+	string cast_op2 =
+	    expression_type(op2).basetype != input_type ? bitcast_glsl(expected_type, op2) : to_unpacked_expression(op2);
 
 	string expr;
 	if (out_type.basetype != input_type)
@@ -4276,7 +4549,7 @@ void CompilerGLSL::emit_texture_op(const Instruction &i)
 	auto op = static_cast<Op>(i.op);
 	uint32_t length = i.length;
 
-	vector<uint32_t> inherited_expressions;
+	SmallVector<uint32_t> inherited_expressions;
 
 	uint32_t result_type = ops[0];
 	uint32_t id = ops[1];
@@ -4426,16 +4699,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 +4873,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 +4974,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);
+				}
 			}
 		}
 	}
@@ -4795,7 +5091,18 @@ void CompilerGLSL::emit_glsl_op(uint32_t result_type, uint32_t id, uint32_t eop,
 		emit_unary_func_op(result_type, id, args[0], "degrees");
 		break;
 	case GLSLstd450Fma:
-		emit_trinary_func_op(result_type, id, args[0], args[1], args[2], "fma");
+		if ((!options.es && options.version < 400) || (options.es && options.version < 320))
+		{
+			auto expr = join(to_enclosed_expression(args[0]), " * ", to_enclosed_expression(args[1]), " + ",
+			                 to_enclosed_expression(args[2]));
+
+			emit_op(result_type, id, expr,
+			        should_forward(args[0]) && should_forward(args[1]) && should_forward(args[2]));
+			for (uint32_t i = 0; i < 3; i++)
+				inherit_expression_dependencies(id, args[i]);
+		}
+		else
+			emit_trinary_func_op(result_type, id, args[0], args[1], args[2], "fma");
 		break;
 	case GLSLstd450Modf:
 		register_call_out_argument(args[1]);
@@ -4807,10 +5114,7 @@ void CompilerGLSL::emit_glsl_op(uint32_t result_type, uint32_t id, uint32_t eop,
 	{
 		forced_temporaries.insert(id);
 		auto &type = get<SPIRType>(result_type);
-		auto &flags = ir.meta[id].decoration.decoration_flags;
-		statement(flags_to_precision_qualifiers_glsl(type, flags), variable_decl(type, to_name(id)), ";");
-		set<SPIRExpression>(id, to_name(id), result_type, true);
-
+		emit_uninitialized_temporary_expression(result_type, id);
 		statement(to_expression(id), ".", to_member_name(type, 0), " = ", "modf(", to_expression(args[0]), ", ",
 		          to_expression(id), ".", to_member_name(type, 1), ");");
 		break;
@@ -4950,10 +5254,7 @@ void CompilerGLSL::emit_glsl_op(uint32_t result_type, uint32_t id, uint32_t eop,
 	{
 		forced_temporaries.insert(id);
 		auto &type = get<SPIRType>(result_type);
-		auto &flags = ir.meta[id].decoration.decoration_flags;
-		statement(flags_to_precision_qualifiers_glsl(type, flags), variable_decl(type, to_name(id)), ";");
-		set<SPIRExpression>(id, to_name(id), result_type, true);
-
+		emit_uninitialized_temporary_expression(result_type, id);
 		statement(to_expression(id), ".", to_member_name(type, 0), " = ", "frexp(", to_expression(args[0]), ", ",
 		          to_expression(id), ".", to_member_name(type, 1), ");");
 		break;
@@ -5033,7 +5334,8 @@ void CompilerGLSL::emit_glsl_op(uint32_t result_type, uint32_t id, uint32_t eop,
 		break;
 
 	case GLSLstd450FindUMsb:
-		emit_unary_func_op_cast(result_type, id, args[0], "findMSB", uint_type, int_type); // findMSB always returns int.
+		emit_unary_func_op_cast(result_type, id, args[0], "findMSB", uint_type,
+		                        int_type); // findMSB always returns int.
 		break;
 
 	// Multisampled varying
@@ -5478,6 +5780,10 @@ case OpGroupNonUniform##op: \
 
 string CompilerGLSL::bitcast_glsl_op(const SPIRType &out_type, const SPIRType &in_type)
 {
+	// OpBitcast can deal with pointers.
+	if (out_type.pointer || in_type.pointer)
+		return type_to_glsl(out_type);
+
 	if (out_type.basetype == in_type.basetype)
 		return "";
 
@@ -5557,9 +5863,9 @@ string CompilerGLSL::bitcast_glsl(const SPIRType &result_type, uint32_t argument
 {
 	auto op = bitcast_glsl_op(result_type, expression_type(argument));
 	if (op.empty())
-		return to_enclosed_expression(argument);
+		return to_enclosed_unpacked_expression(argument);
 	else
-		return join(op, "(", to_expression(argument), ")");
+		return join(op, "(", to_unpacked_expression(argument), ")");
 }
 
 std::string CompilerGLSL::bitcast_expression(SPIRType::BaseType target_type, uint32_t arg)
@@ -5846,6 +6152,21 @@ string CompilerGLSL::access_chain_internal(uint32_t base, const uint32_t *indice
 	// Start traversing type hierarchy at the proper non-pointer types,
 	// but keep type_id referencing the original pointer for use below.
 	uint32_t type_id = expression_type_id(base);
+
+	if (!backend.native_pointers)
+	{
+		if (ptr_chain)
+			SPIRV_CROSS_THROW("Backend does not support native pointers and does not support OpPtrAccessChain.");
+
+		// Wrapped buffer reference pointer types will need to poke into the internal "value" member before
+		// continuing the access chain.
+		if (should_dereference(base))
+		{
+			auto &type = get<SPIRType>(type_id);
+			expr = dereference_expression(type, expr);
+		}
+	}
+
 	const auto *type = &get_pointee_type(type_id);
 
 	bool access_chain_is_arrayed = expr.find_first_of('[') != string::npos;
@@ -5856,19 +6177,34 @@ string CompilerGLSL::access_chain_internal(uint32_t base, const uint32_t *indice
 	bool pending_array_enclose = false;
 	bool dimension_flatten = false;
 
+	const auto append_index = [&](uint32_t index) {
+		expr += "[";
+
+		// If we are indexing into an array of SSBOs or UBOs, we need to index it with a non-uniform qualifier.
+		bool nonuniform_index =
+		    has_decoration(index, DecorationNonUniformEXT) &&
+		    (has_decoration(type->self, DecorationBlock) || has_decoration(type->self, DecorationBufferBlock));
+		if (nonuniform_index)
+		{
+			expr += backend.nonuniform_qualifier;
+			expr += "(";
+		}
+
+		if (index_is_literal)
+			expr += convert_to_string(index);
+		else
+			expr += to_expression(index, register_expression_read);
+
+		if (nonuniform_index)
+			expr += ")";
+
+		expr += "]";
+	};
+
 	for (uint32_t i = 0; i < count; i++)
 	{
 		uint32_t index = indices[i];
 
-		const auto append_index = [&]() {
-			expr += "[";
-			if (index_is_literal)
-				expr += convert_to_string(index);
-			else
-				expr += to_expression(index, register_expression_read);
-			expr += "]";
-		};
-
 		// Pointer chains
 		if (ptr_chain && i == 0)
 		{
@@ -5906,7 +6242,7 @@ string CompilerGLSL::access_chain_internal(uint32_t base, const uint32_t *indice
 			}
 			else
 			{
-				append_index();
+				append_index(index);
 			}
 
 			if (type->basetype == SPIRType::ControlPointArray)
@@ -5953,11 +6289,11 @@ string CompilerGLSL::access_chain_internal(uint32_t base, const uint32_t *indice
 					else if (var->storage == StorageClassOutput)
 						expr = join("gl_out[", to_expression(index, register_expression_read), "].", expr);
 					else
-						append_index();
+						append_index(index);
 					break;
 
 				default:
-					append_index();
+					append_index(index);
 					break;
 				}
 			}
@@ -5987,7 +6323,7 @@ string CompilerGLSL::access_chain_internal(uint32_t base, const uint32_t *indice
 			}
 			else
 			{
-				append_index();
+				append_index(index);
 			}
 
 			type_id = type->parent_type;
@@ -6624,7 +6960,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();
 		}
 	}
 }
@@ -6693,8 +7029,7 @@ void CompilerGLSL::flush_variable_declaration(uint32_t id)
 		{
 			auto &type = get<SPIRType>(var->basetype);
 			auto &flags = ir.meta[id].decoration.decoration_flags;
-			statement(flags_to_precision_qualifiers_glsl(type, flags), variable_decl(type, join("_", id, "_copy")),
-			          ";");
+			statement(flags_to_qualifiers_glsl(type, flags), variable_decl(type, join("_", id, "_copy")), ";");
 		}
 		var->deferred_declaration = false;
 	}
@@ -6953,7 +7288,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)
@@ -7012,6 +7347,10 @@ uint32_t CompilerGLSL::get_integer_width_for_instruction(const Instruction &inst
 
 	switch (instr.op)
 	{
+	case OpSConvert:
+	case OpConvertSToF:
+	case OpUConvert:
+	case OpConvertUToF:
 	case OpIEqual:
 	case OpINotEqual:
 	case OpSLessThan:
@@ -7125,18 +7464,32 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction)
 		// Similar workarounds are required for input arrays in tessellation.
 		unroll_array_from_complex_load(id, ptr, expr);
 
+		auto &type = get<SPIRType>(result_type);
+		if (has_decoration(id, DecorationNonUniformEXT))
+			convert_non_uniform_expression(type, expr);
+
 		if (ptr_expression)
 			ptr_expression->need_transpose = old_need_transpose;
 
 		// By default, suppress usage tracking since using same expression multiple times does not imply any extra work.
 		// However, if we try to load a complex, composite object from a flattened buffer,
 		// we should avoid emitting the same code over and over and lower the result to a temporary.
-		auto &type = get<SPIRType>(result_type);
 		bool usage_tracking = ptr_expression && flattened_buffer_blocks.count(ptr_expression->loaded_from) != 0 &&
 		                      (type.basetype == SPIRType::Struct || (type.columns > 1));
 
-		auto &e = emit_op(result_type, id, expr, forward, !usage_tracking);
-		e.need_transpose = need_transpose;
+		SPIRExpression *e = nullptr;
+		if (!backend.array_is_value_type && !type.array.empty() && !forward)
+		{
+			// Complicated load case where we need to make a copy of ptr, but we cannot, because
+			// it is an array, and our backend does not support arrays as value types.
+			// Emit the temporary, and copy it explicitly.
+			e = &emit_uninitialized_temporary_expression(result_type, id);
+			emit_array_copy(to_expression(id), ptr);
+		}
+		else
+			e = &emit_op(result_type, id, expr, forward, !usage_tracking);
+
+		e->need_transpose = need_transpose;
 		register_read(id, ptr, forward);
 
 		// Pass through whether the result is of a packed type.
@@ -7149,7 +7502,7 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction)
 
 		inherit_expression_dependencies(id, ptr);
 		if (forward)
-			add_implied_read_expression(e, ptr);
+			add_implied_read_expression(*e, ptr);
 		break;
 	}
 
@@ -7225,7 +7578,8 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction)
 		uint32_t result_type = ops[0];
 		uint32_t id = ops[1];
 		auto e = access_chain_internal(ops[2], &ops[3], length - 3, ACCESS_CHAIN_INDEX_IS_LITERAL_BIT, nullptr);
-		set<SPIRExpression>(id, e + ".length()", result_type, true);
+		set<SPIRExpression>(id, join(type_to_glsl(get<SPIRType>(result_type)), "(", e, ".length())"), result_type,
+		                    true);
 		break;
 	}
 
@@ -7267,7 +7621,7 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction)
 			register_impure_function_call();
 
 		string funexpr;
-		vector<string> arglist;
+		SmallVector<string> arglist;
 		funexpr += to_name(func) + "(";
 
 		if (emit_return_value_as_argument)
@@ -7396,10 +7750,7 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction)
 		{
 			// We cannot construct array of arrays because we cannot treat the inputs
 			// as value types. Need to declare the array-of-arrays, and copy in elements one by one.
-			forced_temporaries.insert(id);
-			auto &flags = ir.meta[id].decoration.decoration_flags;
-			statement(flags_to_precision_qualifiers_glsl(out_type, flags), variable_decl(out_type, to_name(id)), ";");
-			set<SPIRExpression>(id, to_name(id), result_type, true);
+			emit_uninitialized_temporary_expression(result_type, id);
 			for (uint32_t i = 0; i < length; i++)
 				emit_array_copy(join(to_expression(id), "[", i, "]"), elems[i]);
 		}
@@ -7633,7 +7984,7 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction)
 			trivial_forward = !expression_is_forwarded(vec0) && !expression_is_forwarded(vec1);
 
 			// Constructor style and shuffling from two different vectors.
-			vector<string> args;
+			SmallVector<string> args;
 			for (uint32_t i = 0; i < length; i++)
 			{
 				if (elems[i] == 0xffffffffu)
@@ -7798,12 +8149,8 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction)
 		uint32_t result_id = ops[1];
 		uint32_t op0 = ops[2];
 		uint32_t op1 = ops[3];
-		forced_temporaries.insert(result_id);
 		auto &type = get<SPIRType>(result_type);
-		auto &flags = ir.meta[result_id].decoration.decoration_flags;
-		statement(flags_to_precision_qualifiers_glsl(type, flags), variable_decl(type, to_name(result_id)), ";");
-		set<SPIRExpression>(result_id, to_name(result_id), result_type, true);
-
+		emit_uninitialized_temporary_expression(result_type, result_id);
 		const char *op = opcode == OpIAddCarry ? "uaddCarry" : "usubBorrow";
 
 		statement(to_expression(result_id), ".", to_member_name(type, 0), " = ", op, "(", to_expression(op0), ", ",
@@ -7825,10 +8172,7 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction)
 		uint32_t op1 = ops[3];
 		forced_temporaries.insert(result_id);
 		auto &type = get<SPIRType>(result_type);
-		auto &flags = ir.meta[result_id].decoration.decoration_flags;
-		statement(flags_to_precision_qualifiers_glsl(type, flags), variable_decl(type, to_name(result_id)), ";");
-		set<SPIRExpression>(result_id, to_name(result_id), result_type, true);
-
+		emit_uninitialized_temporary_expression(result_type, result_id);
 		const char *op = opcode == OpUMulExtended ? "umulExtended" : "imulExtended";
 
 		statement(op, "(", to_expression(op0), ", ", to_expression(op1), ", ", to_expression(result_id), ".",
@@ -8084,12 +8428,45 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction)
 	}
 
 	// Conversion
-	case OpConvertFToU:
-	case OpConvertFToS:
+	case OpSConvert:
 	case OpConvertSToF:
-	case OpConvertUToF:
 	case OpUConvert:
-	case OpSConvert:
+	case OpConvertUToF:
+	{
+		auto input_type = opcode == OpSConvert || opcode == OpConvertSToF ? int_type : uint_type;
+		uint32_t result_type = ops[0];
+		uint32_t id = ops[1];
+
+		auto &type = get<SPIRType>(result_type);
+		auto &arg_type = expression_type(ops[2]);
+		auto func = type_to_glsl_constructor(type);
+
+		// If we're sign-extending or zero-extending, we need to make sure we cast from the correct type.
+		// For truncation, it does not matter, so don't emit useless casts.
+		if (arg_type.width < type.width)
+			emit_unary_func_op_cast(result_type, id, ops[2], func.c_str(), input_type, type.basetype);
+		else
+			emit_unary_func_op(result_type, id, ops[2], func.c_str());
+		break;
+	}
+
+	case OpConvertFToU:
+	case OpConvertFToS:
+	{
+		// Cast to expected arithmetic type, then potentially bitcast away to desired signedness.
+		uint32_t result_type = ops[0];
+		uint32_t id = ops[1];
+		auto &type = get<SPIRType>(result_type);
+		auto expected_type = type;
+		auto &float_type = expression_type(ops[2]);
+		expected_type.basetype =
+		    opcode == OpConvertFToS ? to_signed_basetype(type.width) : to_unsigned_basetype(type.width);
+
+		auto func = type_to_glsl_constructor(expected_type);
+		emit_unary_func_op_cast(result_type, id, ops[2], func.c_str(), float_type.basetype, expected_type.basetype);
+		break;
+	}
+
 	case OpFConvert:
 	{
 		uint32_t result_type = ops[0];
@@ -8293,8 +8670,8 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction)
 		flush_all_atomic_capable_variables();
 		// FIXME: Image?
 		// OpAtomicLoad seems to only be relevant for atomic counters.
+		forced_temporaries.insert(ops[1]);
 		GLSL_UFOP(atomicCounter);
-		register_read(ops[1], ops[2], should_forward(ops[2]));
 		break;
 
 	case OpAtomicStore:
@@ -8334,7 +8711,6 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction)
 		}
 
 		flush_all_atomic_capable_variables();
-		register_read(ops[1], ops[2], should_forward(ops[2]));
 		break;
 	}
 
@@ -8344,7 +8720,6 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction)
 		forced_temporaries.insert(ops[1]);
 		emit_binary_func_op(ops[0], ops[1], ops[2], ops[5], op);
 		flush_all_atomic_capable_variables();
-		register_read(ops[1], ops[2], should_forward(ops[2]));
 		break;
 	}
 
@@ -8355,7 +8730,6 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction)
 		auto expr = join(op, "(", to_expression(ops[2]), ", -", to_enclosed_expression(ops[5]), ")");
 		emit_op(ops[0], ops[1], expr, should_forward(ops[2]) && should_forward(ops[5]));
 		flush_all_atomic_capable_variables();
-		register_read(ops[1], ops[2], should_forward(ops[2]));
 		break;
 	}
 
@@ -8366,7 +8740,6 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction)
 		forced_temporaries.insert(ops[1]);
 		emit_binary_func_op(ops[0], ops[1], ops[2], ops[5], op);
 		flush_all_atomic_capable_variables();
-		register_read(ops[1], ops[2], should_forward(ops[2]));
 		break;
 	}
 
@@ -8377,7 +8750,6 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction)
 		forced_temporaries.insert(ops[1]);
 		emit_binary_func_op(ops[0], ops[1], ops[2], ops[5], op);
 		flush_all_atomic_capable_variables();
-		register_read(ops[1], ops[2], should_forward(ops[2]));
 		break;
 	}
 
@@ -8387,7 +8759,6 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction)
 		forced_temporaries.insert(ops[1]);
 		emit_binary_func_op(ops[0], ops[1], ops[2], ops[5], op);
 		flush_all_atomic_capable_variables();
-		register_read(ops[1], ops[2], should_forward(ops[2]));
 		break;
 	}
 
@@ -8397,7 +8768,6 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction)
 		forced_temporaries.insert(ops[1]);
 		emit_binary_func_op(ops[0], ops[1], ops[2], ops[5], op);
 		flush_all_atomic_capable_variables();
-		register_read(ops[1], ops[2], should_forward(ops[2]));
 		break;
 	}
 
@@ -8407,7 +8777,6 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction)
 		forced_temporaries.insert(ops[1]);
 		emit_binary_func_op(ops[0], ops[1], ops[2], ops[5], op);
 		flush_all_atomic_capable_variables();
-		register_read(ops[1], ops[2], should_forward(ops[2]));
 		break;
 	}
 
@@ -8543,7 +8912,7 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction)
 			if (flags.get(DecorationNonReadable))
 			{
 				flags.clear(DecorationNonReadable);
-				force_recompile = true;
+				force_recompile();
 			}
 		}
 
@@ -8691,7 +9060,7 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction)
 			if (flags.get(DecorationNonWritable))
 			{
 				flags.clear(DecorationNonWritable);
-				force_recompile = true;
+				force_recompile();
 			}
 		}
 
@@ -9168,6 +9537,33 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction)
 		statement("executeCallableNV(", to_expression(ops[0]), ", ", to_expression(ops[1]), ");");
 		break;
 
+	case OpConvertUToPtr:
+	{
+		auto &type = get<SPIRType>(ops[0]);
+		if (type.storage != StorageClassPhysicalStorageBufferEXT)
+			SPIRV_CROSS_THROW("Only StorageClassPhysicalStorageBufferEXT is supported by OpConvertUToPtr.");
+
+		auto op = type_to_glsl(type);
+		emit_unary_func_op(ops[0], ops[1], ops[2], op.c_str());
+		break;
+	}
+
+	case OpConvertPtrToU:
+	{
+		auto &type = get<SPIRType>(ops[0]);
+		auto &ptr_type = expression_type(ops[2]);
+		if (ptr_type.storage != StorageClassPhysicalStorageBufferEXT)
+			SPIRV_CROSS_THROW("Only StorageClassPhysicalStorageBufferEXT is supported by OpConvertPtrToU.");
+
+		auto op = type_to_glsl(type);
+		emit_unary_func_op(ops[0], ops[1], ops[2], op.c_str());
+		break;
+	}
+
+	case OpUndef:
+		// Undefined value has been declared.
+		break;
+
 	default:
 		statement("// unimplemented op ", instruction.op);
 		break;
@@ -9182,7 +9578,7 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction)
 // access to shader input content from within a function (eg. Metal). Each additional
 // function args uses the name of the global variable. Function nesting will modify the
 // functions and function calls all the way up the nesting chain.
-void CompilerGLSL::append_global_func_args(const SPIRFunction &func, uint32_t index, vector<string> &arglist)
+void CompilerGLSL::append_global_func_args(const SPIRFunction &func, uint32_t index, SmallVector<string> &arglist)
 {
 	auto &args = func.arguments;
 	uint32_t arg_cnt = uint32_t(args.size());
@@ -9321,13 +9717,16 @@ void CompilerGLSL::emit_struct_member(const SPIRType &type, uint32_t member_type
 	if (is_block)
 		qualifiers = to_interpolation_qualifiers(memberflags);
 
-	statement(layout_for_member(type, index), qualifiers, qualifier,
-	          flags_to_precision_qualifiers_glsl(membertype, memberflags),
+	statement(layout_for_member(type, index), qualifiers, qualifier, flags_to_qualifiers_glsl(membertype, memberflags),
 	          variable_decl(membertype, to_member_name(type, index)), ";");
 }
 
-const char *CompilerGLSL::flags_to_precision_qualifiers_glsl(const SPIRType &type, const Bitset &flags)
+const char *CompilerGLSL::flags_to_qualifiers_glsl(const SPIRType &type, const Bitset &flags)
 {
+	// GL_EXT_buffer_reference variables can be marked as restrict.
+	if (flags.get(DecorationRestrictPointerEXT))
+		return "restrict ";
+
 	// Structs do not have precision qualifiers, neither do doubles (desktop only anyways, so no mediump/highp).
 	if (type.basetype != SPIRType::Float && type.basetype != SPIRType::Int && type.basetype != SPIRType::UInt &&
 	    type.basetype != SPIRType::Image && type.basetype != SPIRType::SampledImage &&
@@ -9380,7 +9779,7 @@ const char *CompilerGLSL::flags_to_precision_qualifiers_glsl(const SPIRType &typ
 
 const char *CompilerGLSL::to_precision_qualifiers_glsl(uint32_t id)
 {
-	return flags_to_precision_qualifiers_glsl(expression_type(id), ir.meta[id].decoration.decoration_flags);
+	return flags_to_qualifiers_glsl(expression_type(id), ir.meta[id].decoration.decoration_flags);
 }
 
 string CompilerGLSL::to_qualifiers_glsl(uint32_t id)
@@ -9527,7 +9926,7 @@ string CompilerGLSL::to_array_size(const SPIRType &type, uint32_t index)
 		return to_expression(size);
 	else if (size)
 		return convert_to_string(size);
-	else if (!backend.flexible_member_array_supported)
+	else if (!backend.unsized_array_supported)
 	{
 		// For runtime-sized arrays, we can work around
 		// lack of standard support for this by simply having
@@ -9543,6 +9942,12 @@ string CompilerGLSL::to_array_size(const SPIRType &type, uint32_t index)
 
 string CompilerGLSL::type_to_array_glsl(const SPIRType &type)
 {
+	if (type.pointer && type.storage == StorageClassPhysicalStorageBufferEXT && type.basetype != SPIRType::Struct)
+	{
+		// We are using a wrapped pointer type, and we should not emit any array declarations here.
+		return "";
+	}
+
 	if (type.array.empty())
 		return "";
 
@@ -9696,7 +10101,20 @@ string CompilerGLSL::type_to_glsl_constructor(const SPIRType &type)
 // depend on a specific object's use of that type.
 string CompilerGLSL::type_to_glsl(const SPIRType &type, uint32_t id)
 {
-	// Ignore the pointer type since GLSL doesn't have pointers.
+	if (type.pointer && type.storage == StorageClassPhysicalStorageBufferEXT && type.basetype != SPIRType::Struct)
+	{
+		// Need to create a magic type name which compacts the entire type information.
+		string name = type_to_glsl(get_pointee_type(type));
+		for (size_t i = 0; i < type.array.size(); i++)
+		{
+			if (type.array_size_literal[i])
+				name += join(type.array[i], "_");
+			else
+				name += join("id", type.array[i], "_");
+		}
+		name += "Pointer";
+		return name;
+	}
 
 	switch (type.basetype)
 	{
@@ -9890,7 +10308,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 +10347,7 @@ bool CompilerGLSL::check_atomic_image(uint32_t id)
 			{
 				flags.clear(DecorationNonWritable);
 				flags.clear(DecorationNonReadable);
-				force_recompile = true;
+				force_recompile();
 			}
 		}
 		return true;
@@ -10003,7 +10421,7 @@ void CompilerGLSL::emit_function_prototype(SPIRFunction &func, const Bitset &ret
 	string decl;
 
 	auto &type = get<SPIRType>(func.return_type);
-	decl += flags_to_precision_qualifiers_glsl(type, return_flags);
+	decl += flags_to_qualifiers_glsl(type, return_flags);
 	decl += type_to_glsl(type);
 	decl += type_to_array_glsl(type);
 	decl += " ";
@@ -10017,7 +10435,7 @@ void CompilerGLSL::emit_function_prototype(SPIRFunction &func, const Bitset &ret
 		decl += to_name(func.self);
 
 	decl += "(";
-	vector<string> arglist;
+	SmallVector<string> arglist;
 	for (auto &arg : func.arguments)
 	{
 		// Do not pass in separate images or samplers if we're remapping
@@ -10237,7 +10655,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 +10765,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;
 			}
 
@@ -10475,7 +10893,7 @@ string CompilerGLSL::emit_continue_block(uint32_t continue_block, bool follow_tr
 	// if we have to emit temporaries.
 	current_continue_block = block;
 
-	vector<string> statements;
+	SmallVector<string> statements;
 
 	// Capture all statements into our list.
 	auto *old = redirect_statement;
@@ -10705,7 +11123,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 +11199,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;
 		}
@@ -10798,7 +11216,7 @@ void CompilerGLSL::flush_undeclared_variables(SPIRBlock &block)
 		flush_variable_declaration(v);
 }
 
-void CompilerGLSL::emit_hoisted_temporaries(vector<pair<uint32_t, uint32_t>> &temporaries)
+void CompilerGLSL::emit_hoisted_temporaries(SmallVector<pair<uint32_t, uint32_t>> &temporaries)
 {
 	// If we need to force temporaries for certain IDs due to continue blocks, do it before starting loop header.
 	// Need to sort these to ensure that reference output is stable.
@@ -10810,7 +11228,7 @@ void CompilerGLSL::emit_hoisted_temporaries(vector<pair<uint32_t, uint32_t>> &te
 		add_local_variable_name(tmp.second);
 		auto &flags = ir.meta[tmp.second].decoration.decoration_flags;
 		auto &type = get<SPIRType>(tmp.first);
-		statement(flags_to_precision_qualifiers_glsl(type, flags), variable_decl(type, to_name(tmp.second)), ";");
+		statement(flags_to_qualifiers_glsl(type, flags), variable_decl(type, to_name(tmp.second)), ";");
 
 		hoisted_temporaries.insert(tmp.second);
 		forced_temporaries.insert(tmp.second);
@@ -10928,7 +11346,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();
@@ -11173,7 +11591,7 @@ void CompilerGLSL::emit_block_chain(SPIRBlock &block)
 			assert(block.merge == SPIRBlock::MergeSelection);
 			branch_to_continue(block.self, block.next_block);
 		}
-		else
+		else if (block.self != block.next_block)
 			emit_block_chain(get<SPIRBlock>(block.next_block));
 	}
 
@@ -11188,12 +11606,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.
@@ -11347,8 +11766,7 @@ void CompilerGLSL::unroll_array_from_complex_load(uint32_t target_id, uint32_t s
 	}
 }
 
-void CompilerGLSL::bitcast_from_builtin_load(uint32_t source_id, std::string &expr,
-                                             const spirv_cross::SPIRType &expr_type)
+void CompilerGLSL::bitcast_from_builtin_load(uint32_t source_id, std::string &expr, const SPIRType &expr_type)
 {
 	auto *var = maybe_get_backing_variable(source_id);
 	if (var)
@@ -11395,8 +11813,7 @@ void CompilerGLSL::bitcast_from_builtin_load(uint32_t source_id, std::string &ex
 		expr = bitcast_expression(expr_type, expected_type, expr);
 }
 
-void CompilerGLSL::bitcast_to_builtin_store(uint32_t target_id, std::string &expr,
-                                            const spirv_cross::SPIRType &expr_type)
+void CompilerGLSL::bitcast_to_builtin_store(uint32_t target_id, std::string &expr, const SPIRType &expr_type)
 {
 	// Only interested in standalone builtin variables.
 	if (!has_decoration(target_id, DecorationBuiltIn))
@@ -11426,6 +11843,38 @@ void CompilerGLSL::bitcast_to_builtin_store(uint32_t target_id, std::string &exp
 	}
 }
 
+void CompilerGLSL::convert_non_uniform_expression(const SPIRType &type, std::string &expr)
+{
+	if (*backend.nonuniform_qualifier == '\0')
+		return;
+
+	// Handle SPV_EXT_descriptor_indexing.
+	if (type.basetype == SPIRType::Sampler || type.basetype == SPIRType::SampledImage ||
+	    type.basetype == SPIRType::Image)
+	{
+		// The image/sampler ID must be declared as non-uniform.
+		// However, it is not legal GLSL to have
+		// nonuniformEXT(samplers[index]), so we must move the nonuniform qualifier
+		// to the array indexing, like
+		// samplers[nonuniformEXT(index)].
+		// While the access chain will generally be nonuniformEXT, it's not necessarily so,
+		// so we might have to fixup the OpLoad-ed expression late.
+
+		auto start_array_index = expr.find_first_of('[');
+		auto end_array_index = expr.find_last_of(']');
+		// Doesn't really make sense to declare a non-arrayed image with nonuniformEXT, but there's
+		// nothing we can do here to express that.
+		if (start_array_index == string::npos || end_array_index == string::npos || end_array_index < start_array_index)
+			return;
+
+		start_array_index++;
+
+		expr = join(expr.substr(0, start_array_index), backend.nonuniform_qualifier, "(",
+		            expr.substr(start_array_index, end_array_index - start_array_index), ")",
+		            expr.substr(end_array_index, string::npos));
+	}
+}
+
 void CompilerGLSL::emit_block_hints(const SPIRBlock &)
 {
 }
diff --git a/src/3rdparty/SPIRV-Cross/spirv_glsl.hpp b/src/3rdparty/SPIRV-Cross/spirv_glsl.hpp
index 33e3547..184bbbd 100644
--- a/src/3rdparty/SPIRV-Cross/spirv_glsl.hpp
+++ b/src/3rdparty/SPIRV-Cross/spirv_glsl.hpp
@@ -17,14 +17,13 @@
 #ifndef SPIRV_CROSS_GLSL_HPP
 #define SPIRV_CROSS_GLSL_HPP
 
-#include "spirv_cross.hpp"
 #include "GLSL.std.450.h"
-#include <sstream>
+#include "spirv_cross.hpp"
 #include <unordered_map>
 #include <unordered_set>
 #include <utility>
 
-namespace spirv_cross
+namespace SPIRV_CROSS_NAMESPACE
 {
 enum PlsFormat
 {
@@ -100,6 +99,10 @@ public:
 		// In non-Vulkan GLSL, emit push constant blocks as UBOs rather than plain uniforms.
 		bool emit_push_constant_as_uniform_buffer = false;
 
+		// Always emit uniform blocks as plain uniforms, regardless of the GLSL version, even when UBOs are supported.
+		// Does not apply to shader storage or push constant blocks.
+		bool emit_uniform_buffer_as_plain_uniforms = false;
+
 		enum Precision
 		{
 			DontCare,
@@ -142,7 +145,7 @@ public:
 	}
 
 	explicit CompilerGLSL(std::vector<uint32_t> spirv_)
-	    : Compiler(move(spirv_))
+	    : Compiler(std::move(spirv_))
 	{
 		init();
 	}
@@ -230,7 +233,7 @@ protected:
 	virtual void emit_spv_amd_gcn_shader_op(uint32_t result_type, uint32_t result_id, uint32_t op, const uint32_t *args,
 	                                        uint32_t count);
 	virtual void emit_header();
-	void build_workgroup_size(std::vector<std::string> &arguments, const SpecializationConstant &x,
+	void build_workgroup_size(SmallVector<std::string> &arguments, const SpecializationConstant &x,
 	                          const SpecializationConstant &y, const SpecializationConstant &z);
 
 	virtual void emit_sampled_image_op(uint32_t result_type, uint32_t result_id, uint32_t image_id, uint32_t samp_id);
@@ -260,19 +263,19 @@ protected:
 	virtual void emit_uniform(const SPIRVariable &var);
 	virtual std::string unpack_expression_type(std::string expr_str, const SPIRType &type, uint32_t packed_type_id);
 
-	std::unique_ptr<std::ostringstream> buffer;
+	StringStream<> buffer;
 
 	template <typename T>
 	inline void statement_inner(T &&t)
 	{
-		(*buffer) << std::forward<T>(t);
+		buffer << std::forward<T>(t);
 		statement_count++;
 	}
 
 	template <typename T, typename... Ts>
 	inline void statement_inner(T &&t, Ts &&... ts)
 	{
-		(*buffer) << std::forward<T>(t);
+		buffer << std::forward<T>(t);
 		statement_count++;
 		statement_inner(std::forward<Ts>(ts)...);
 	}
@@ -280,7 +283,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,13 +292,16 @@ protected:
 		}
 
 		if (redirect_statement)
+		{
 			redirect_statement->push_back(join(std::forward<Ts>(ts)...));
+			statement_count++;
+		}
 		else
 		{
 			for (uint32_t i = 0; i < indent; i++)
-				(*buffer) << "    ";
+				buffer << "    ";
 			statement_inner(std::forward<Ts>(ts)...);
-			(*buffer) << '\n';
+			buffer << '\n';
 		}
 	}
 
@@ -311,7 +317,7 @@ protected:
 	// Used for implementing continue blocks where
 	// we want to obtain a list of statements we can merge
 	// on a single line separated by comma.
-	std::vector<std::string> *redirect_statement = nullptr;
+	SmallVector<std::string> *redirect_statement = nullptr;
 	const SPIRBlock *current_continue_block = nullptr;
 
 	void begin_scope();
@@ -370,9 +376,10 @@ protected:
 		const char *basic_uint16_type = "uint16_t";
 		const char *int16_t_literal_suffix = "s";
 		const char *uint16_t_literal_suffix = "us";
+		const char *nonuniform_qualifier = "nonuniformEXT";
 		bool swizzle_is_function = false;
 		bool shared_is_implied = false;
-		bool flexible_member_array_supported = true;
+		bool unsized_array_supported = true;
 		bool explicit_struct_type = false;
 		bool use_initializer_list = false;
 		bool use_typed_initializer_list = false;
@@ -390,11 +397,13 @@ protected:
 		bool supports_empty_struct = false;
 		bool array_is_value_type = true;
 		bool comparison_image_samples_scalar = false;
+		bool native_pointers = false;
 	} backend;
 
 	void emit_struct(SPIRType &type);
 	void emit_resources();
 	void emit_buffer_block_native(const SPIRVariable &var);
+	void emit_buffer_reference_block(SPIRType &type, bool forward_declaration);
 	void emit_buffer_block_legacy(const SPIRVariable &var);
 	void emit_buffer_block_flattened(const SPIRVariable &type);
 	void emit_declared_builtin_block(spv::StorageClass storage, spv::ExecutionModel model);
@@ -403,7 +412,7 @@ protected:
 	void emit_interface_block(const SPIRVariable &type);
 	void emit_flattened_io_block(const SPIRVariable &var, const char *qual);
 	void emit_block_chain(SPIRBlock &block);
-	void emit_hoisted_temporaries(std::vector<std::pair<uint32_t, uint32_t>> &temporaries);
+	void emit_hoisted_temporaries(SmallVector<std::pair<uint32_t, uint32_t>> &temporaries);
 	std::string constant_value_macro_name(uint32_t id);
 	void emit_constant(const SPIRConstant &constant);
 	void emit_specialization_constant_op(const SPIRConstantOp &constant);
@@ -434,8 +443,8 @@ protected:
 	                             SPIRType::BaseType input_type, SPIRType::BaseType expected_result_type);
 	void emit_binary_func_op_cast(uint32_t result_type, uint32_t result_id, uint32_t op0, uint32_t op1, const char *op,
 	                              SPIRType::BaseType input_type, bool skip_cast_if_equal_type);
-	void emit_trinary_func_op_cast(uint32_t result_type, uint32_t result_id, uint32_t op0, uint32_t op1, uint32_t op2, const char *op,
-	                               SPIRType::BaseType input_type);
+	void emit_trinary_func_op_cast(uint32_t result_type, uint32_t result_id, uint32_t op0, uint32_t op1, uint32_t op2,
+	                               const char *op, SPIRType::BaseType input_type);
 
 	void emit_unary_func_op(uint32_t result_type, uint32_t result_id, uint32_t op0, const char *op);
 	void emit_unrolled_unary_op(uint32_t result_type, uint32_t result_id, uint32_t operand, const char *op);
@@ -481,7 +490,9 @@ protected:
 	const char *index_to_swizzle(uint32_t index);
 	std::string remap_swizzle(const SPIRType &result_type, uint32_t input_components, const std::string &expr);
 	std::string declare_temporary(uint32_t type, uint32_t id);
-	void append_global_func_args(const SPIRFunction &func, uint32_t index, std::vector<std::string> &arglist);
+	void emit_uninitialized_temporary(uint32_t type, uint32_t id);
+	SPIRExpression &emit_uninitialized_temporary_expression(uint32_t type, uint32_t id);
+	void append_global_func_args(const SPIRFunction &func, uint32_t index, SmallVector<std::string> &arglist);
 	std::string to_expression(uint32_t id, bool register_expression_read = true);
 	std::string to_enclosed_expression(uint32_t id, bool register_expression_read = true);
 	std::string to_unpacked_expression(uint32_t id, bool register_expression_read = true);
@@ -491,7 +502,7 @@ protected:
 	std::string to_enclosed_pointer_expression(uint32_t id, bool register_expression_read = true);
 	std::string to_extract_component_expression(uint32_t id, uint32_t index);
 	std::string enclose_expression(const std::string &expr);
-	std::string dereference_expression(const std::string &expr);
+	std::string dereference_expression(const SPIRType &expression_type, const std::string &expr);
 	std::string address_of_expression(const std::string &expr);
 	void strip_enclosed_expression(std::string &expr);
 	std::string to_member_name(const SPIRType &type, uint32_t index);
@@ -501,7 +512,7 @@ protected:
 	virtual std::string to_qualifiers_glsl(uint32_t id);
 	const char *to_precision_qualifiers_glsl(uint32_t id);
 	virtual const char *to_storage_qualifiers_glsl(const SPIRVariable &var);
-	const char *flags_to_precision_qualifiers_glsl(const SPIRType &type, const Bitset &flags);
+	const char *flags_to_qualifiers_glsl(const SPIRType &type, const Bitset &flags);
 	const char *format_to_glsl(spv::ImageFormat format);
 	virtual std::string layout_for_member(const SPIRType &type, uint32_t index);
 	virtual std::string to_interpolation_qualifiers(const Bitset &flags);
@@ -514,6 +525,8 @@ protected:
 
 	bool buffer_is_packing_standard(const SPIRType &type, BufferPackingStandard packing, uint32_t start_offset = 0,
 	                                uint32_t end_offset = ~(0u));
+	std::string buffer_to_packing_standard(const SPIRType &type, bool enable_std430);
+
 	uint32_t type_to_packed_base_size(const SPIRType &type, BufferPackingStandard packing);
 	uint32_t type_to_packed_alignment(const SPIRType &type, const Bitset &flags, BufferPackingStandard packing);
 	uint32_t type_to_packed_array_stride(const SPIRType &type, const Bitset &flags, BufferPackingStandard packing);
@@ -557,15 +570,15 @@ protected:
 	std::unordered_map<uint32_t, uint32_t> expression_usage_counts;
 	void track_expression_read(uint32_t id);
 
-	std::vector<std::string> forced_extensions;
-	std::vector<std::string> header_lines;
+	SmallVector<std::string> forced_extensions;
+	SmallVector<std::string> header_lines;
 
 	// Used when expressions emit extra opcodes with their own unique IDs,
 	// and we need to reuse the IDs across recompilation loops.
 	// Currently used by NMin/Max/Clamp implementations.
 	std::unordered_map<uint32_t, uint32_t> extra_sub_expressions;
 
-	uint32_t statement_count;
+	uint32_t statement_count = 0;
 
 	inline bool is_legacy() const
 	{
@@ -633,6 +646,7 @@ protected:
 	virtual void bitcast_to_builtin_store(uint32_t target_id, std::string &expr, const SPIRType &expr_type);
 	virtual void bitcast_from_builtin_load(uint32_t source_id, std::string &expr, const SPIRType &expr_type);
 	void unroll_array_from_complex_load(uint32_t target_id, uint32_t source_id, std::string &expr);
+	void convert_non_uniform_expression(const SPIRType &type, std::string &expr);
 
 	void handle_store_to_invariant_variable(uint32_t store_id, uint32_t value_id);
 	void disallow_forwarding_in_expression_chain(const SPIRExpression &expr);
@@ -650,6 +664,6 @@ protected:
 private:
 	void init();
 };
-} // namespace spirv_cross
+} // namespace SPIRV_CROSS_NAMESPACE
 
 #endif
diff --git a/src/3rdparty/SPIRV-Cross/spirv_hlsl.cpp b/src/3rdparty/SPIRV-Cross/spirv_hlsl.cpp
index 3f6b627..46613c5 100644
--- a/src/3rdparty/SPIRV-Cross/spirv_hlsl.cpp
+++ b/src/3rdparty/SPIRV-Cross/spirv_hlsl.cpp
@@ -20,7 +20,7 @@
 #include <assert.h>
 
 using namespace spv;
-using namespace spirv_cross;
+using namespace SPIRV_CROSS_NAMESPACE;
 using namespace std;
 
 static unsigned image_format_to_components(ImageFormat fmt)
@@ -723,17 +723,25 @@ string CompilerHLSL::to_interpolation_qualifiers(const Bitset &flags)
 	return res;
 }
 
-std::string CompilerHLSL::to_semantic(uint32_t vertex_location)
+std::string CompilerHLSL::to_semantic(uint32_t location, ExecutionModel em, StorageClass sc)
 {
-	for (auto &attribute : remap_vertex_attributes)
-		if (attribute.location == vertex_location)
-			return attribute.semantic;
+	if (em == ExecutionModelVertex && sc == StorageClassInput)
+	{
+		// We have a vertex attribute - we should look at remapping it if the user provided
+		// vertex attribute hints.
+		for (auto &attribute : remap_vertex_attributes)
+			if (attribute.location == location)
+				return attribute.semantic;
+	}
 
-	return join("TEXCOORD", vertex_location);
+	// Not a vertex attribute, or no remap_vertex_attributes entry.
+	return join("TEXCOORD", location);
 }
 
 void CompilerHLSL::emit_io_block(const SPIRVariable &var)
 {
+	auto &execution = get_entry_point();
+
 	auto &type = get<SPIRType>(var.basetype);
 	add_resource_name(type.self);
 
@@ -749,7 +757,7 @@ void CompilerHLSL::emit_io_block(const SPIRVariable &var)
 		if (has_member_decoration(type.self, i, DecorationLocation))
 		{
 			uint32_t location = get_member_decoration(type.self, i, DecorationLocation);
-			semantic = join(" : ", to_semantic(location));
+			semantic = join(" : ", to_semantic(location, execution.model, var.storage));
 		}
 		else
 		{
@@ -757,7 +765,7 @@ void CompilerHLSL::emit_io_block(const SPIRVariable &var)
 			// There could be a conflict if the block members partially specialize the locations.
 			// It is unclear how SPIR-V deals with this. Assume this does not happen for now.
 			uint32_t location = base_location + i;
-			semantic = join(" : ", to_semantic(location));
+			semantic = join(" : ", to_semantic(location, execution.model, var.storage));
 		}
 
 		add_member_name(type, i);
@@ -820,7 +828,7 @@ void CompilerHLSL::emit_interface_block_in_struct(const SPIRVariable &var, unord
 			location_number = get_vacant_location();
 
 		// Allow semantic remap if specified.
-		auto semantic = to_semantic(location_number);
+		auto semantic = to_semantic(location_number, execution.model, var.storage);
 
 		if (need_matrix_unroll && type.columns > 1)
 		{
@@ -1210,8 +1218,8 @@ void CompilerHLSL::emit_resources()
 	require_output = false;
 	unordered_set<uint32_t> active_inputs;
 	unordered_set<uint32_t> active_outputs;
-	vector<SPIRVariable *> input_variables;
-	vector<SPIRVariable *> output_variables;
+	SmallVector<SPIRVariable *> input_variables;
+	SmallVector<SPIRVariable *> output_variables;
 	ir.for_each_typed_id<SPIRVariable>([&](uint32_t, SPIRVariable &var) {
 		auto &type = this->get<SPIRType>(var.basetype);
 		bool block = ir.meta[type.self].decoration.decoration_flags.get(DecorationBlock);
@@ -1998,7 +2006,7 @@ void CompilerHLSL::emit_function_prototype(SPIRFunction &func, const Bitset &ret
 	auto &type = get<SPIRType>(func.return_type);
 	if (type.array.empty())
 	{
-		decl += flags_to_precision_qualifiers_glsl(type, return_flags);
+		decl += flags_to_qualifiers_glsl(type, return_flags);
 		decl += type_to_glsl(type);
 		decl += " ";
 	}
@@ -2024,7 +2032,7 @@ void CompilerHLSL::emit_function_prototype(SPIRFunction &func, const Bitset &ret
 		decl += to_name(func.self);
 
 	decl += "(";
-	vector<string> arglist;
+	SmallVector<string> arglist;
 
 	if (!type.array.empty())
 	{
@@ -2092,7 +2100,7 @@ void CompilerHLSL::emit_function_prototype(SPIRFunction &func, const Bitset &ret
 
 void CompilerHLSL::emit_hlsl_entry_point()
 {
-	vector<string> arguments;
+	SmallVector<string> arguments;
 
 	if (require_input)
 		arguments.push_back("SPIRV_Cross_Input stage_input");
@@ -2425,7 +2433,7 @@ void CompilerHLSL::emit_texture_op(const Instruction &i)
 	auto op = static_cast<Op>(i.op);
 	uint32_t length = i.length;
 
-	vector<uint32_t> inherited_expressions;
+	SmallVector<uint32_t> inherited_expressions;
 
 	uint32_t result_type = ops[0];
 	uint32_t id = ops[1];
@@ -2716,7 +2724,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 +2735,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 +2745,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 +2755,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 +3036,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 +3045,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 +3107,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 +3116,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 +3125,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 +3134,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 +3143,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 +3152,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 +3161,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 +3170,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 +3179,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 +3188,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 +3217,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 +3225,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 +3233,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");
@@ -3707,7 +3713,6 @@ void CompilerHLSL::emit_atomic(const uint32_t *ops, uint32_t length, spv::Op op)
 	auto expr = bitcast_expression(type, expr_type, to_name(id));
 	set<SPIRExpression>(id, expr, result_type, true);
 	flush_all_atomic_capable_variables();
-	register_read(ops[1], ops[2], should_forward(ops[2]));
 }
 
 void CompilerHLSL::emit_subgroup_op(const Instruction &i)
@@ -3949,7 +3954,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 +4469,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 +4490,7 @@ void CompilerHLSL::emit_instruction(const Instruction &instruction)
 		if (!requires_bitfield_extract)
 		{
 			requires_bitfield_extract = true;
-			force_recompile = true;
+			force_recompile();
 		}
 
 		if (opcode == OpBitFieldSExtract)
@@ -4503,6 +4508,25 @@ void CompilerHLSL::emit_instruction(const Instruction &instruction)
 		HLSL_UFOP(reversebits);
 		break;
 
+	case OpArrayLength:
+	{
+		auto *var = maybe_get<SPIRVariable>(ops[2]);
+		if (!var)
+			SPIRV_CROSS_THROW("Array length must point directly to an SSBO block.");
+
+		auto &type = get<SPIRType>(var->basetype);
+		if (!has_decoration(type.self, DecorationBlock) && !has_decoration(type.self, DecorationBufferBlock))
+			SPIRV_CROSS_THROW("Array length expression must point to a block type.");
+
+		// This must be 32-bit uint, so we're good to go.
+		emit_uninitialized_temporary_expression(ops[0], ops[1]);
+		statement(to_expression(ops[2]), ".GetDimensions(", to_expression(ops[1]), ");");
+		uint32_t offset = type_struct_member_offset(type, ops[3]);
+		uint32_t stride = type_struct_member_array_stride(type, ops[3]);
+		statement(to_expression(ops[1]), " = (", to_expression(ops[1]), " - ", offset, ") / ", stride, ";");
+		break;
+	}
+
 	default:
 		CompilerGLSL::emit_instruction(instruction);
 		break;
@@ -4562,12 +4586,12 @@ 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;
 	}
 }
 
-void CompilerHLSL::set_root_constant_layouts(vector<RootConstants> layout)
+void CompilerHLSL::set_root_constant_layouts(std::vector<RootConstants> layout)
 {
 	root_constants_layout = move(layout);
 }
@@ -4623,6 +4647,28 @@ uint32_t CompilerHLSL::remap_num_workgroups_builtin()
 	return variable_id;
 }
 
+void CompilerHLSL::validate_shader_model()
+{
+	// Check for nonuniform qualifier.
+	// Instead of looping over all decorations to find this, just look at capabilities.
+	for (auto &cap : ir.declared_capabilities)
+	{
+		switch (cap)
+		{
+		case CapabilityShaderNonUniformEXT:
+		case CapabilityRuntimeDescriptorArrayEXT:
+			if (hlsl_options.shader_model < 51)
+				SPIRV_CROSS_THROW(
+				    "Shader model 5.1 or higher is required to use bindless resources or NonUniformResourceIndex.");
+		default:
+			break;
+		}
+	}
+
+	if (ir.addressing_model != AddressingModelLogical)
+		SPIRV_CROSS_THROW("Only Logical addressing model can be used with HLSL.");
+}
+
 string CompilerHLSL::compile()
 {
 	// Do not deal with ES-isms like precision, older extensions and such.
@@ -4639,7 +4685,7 @@ string CompilerHLSL::compile()
 	backend.basic_uint_type = "uint";
 	backend.swizzle_is_function = false;
 	backend.shared_is_implied = true;
-	backend.flexible_member_array_supported = false;
+	backend.unsized_array_supported = true;
 	backend.explicit_struct_type = false;
 	backend.use_initializer_list = true;
 	backend.use_constructor_splatting = false;
@@ -4648,8 +4694,10 @@ string CompilerHLSL::compile()
 	backend.can_declare_struct_inline = false;
 	backend.can_declare_arrays_inline = false;
 	backend.can_return_array = false;
+	backend.nonuniform_qualifier = "NonUniformResourceIndex";
 
 	build_function_control_flow_graphs_and_analyze();
+	validate_shader_model();
 	update_active_builtins();
 	analyze_image_and_sampler_usage();
 
@@ -4666,7 +4714,7 @@ string CompilerHLSL::compile()
 		reset();
 
 		// Move constructor for this type is broken on GCC 4.9 ...
-		buffer = unique_ptr<ostringstream>(new ostringstream());
+		buffer.reset();
 
 		emit_header();
 		emit_resources();
@@ -4675,12 +4723,12 @@ 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";
 
-	return buffer->str();
+	return buffer.str();
 }
 
 void CompilerHLSL::emit_block_hints(const SPIRBlock &block)
diff --git a/src/3rdparty/SPIRV-Cross/spirv_hlsl.hpp b/src/3rdparty/SPIRV-Cross/spirv_hlsl.hpp
index 12b8ae1..d96c911 100644
--- a/src/3rdparty/SPIRV-Cross/spirv_hlsl.hpp
+++ b/src/3rdparty/SPIRV-Cross/spirv_hlsl.hpp
@@ -19,9 +19,8 @@
 
 #include "spirv_glsl.hpp"
 #include <utility>
-#include <vector>
 
-namespace spirv_cross
+namespace SPIRV_CROSS_NAMESPACE
 {
 // Interface which remaps vertex inputs to a fixed semantic name to make linking easier.
 struct HLSLVertexAttributeRemap
@@ -63,7 +62,7 @@ public:
 	};
 
 	explicit CompilerHLSL(std::vector<uint32_t> spirv_)
-	    : CompilerGLSL(move(spirv_))
+	    : CompilerGLSL(std::move(spirv_))
 	{
 	}
 
@@ -209,19 +208,21 @@ private:
 	void emit_builtin_variables();
 	bool require_output = false;
 	bool require_input = false;
-	std::vector<HLSLVertexAttributeRemap> remap_vertex_attributes;
+	SmallVector<HLSLVertexAttributeRemap> remap_vertex_attributes;
 
 	uint32_t type_to_consumed_locations(const SPIRType &type) const;
 
 	void emit_io_block(const SPIRVariable &var);
-	std::string to_semantic(uint32_t vertex_location);
+	std::string to_semantic(uint32_t location, spv::ExecutionModel em, spv::StorageClass sc);
 
 	uint32_t num_workgroups_builtin = 0;
 
 	// Custom root constant layout, which should be emitted
 	// when translating push constant ranges.
 	std::vector<RootConstants> root_constants_layout;
+
+	void validate_shader_model();
 };
-} // namespace spirv_cross
+} // namespace SPIRV_CROSS_NAMESPACE
 
 #endif
diff --git a/src/3rdparty/SPIRV-Cross/spirv_msl.cpp b/src/3rdparty/SPIRV-Cross/spirv_msl.cpp
index 41a3aaa..4a4f77a 100644
--- a/src/3rdparty/SPIRV-Cross/spirv_msl.cpp
+++ b/src/3rdparty/SPIRV-Cross/spirv_msl.cpp
@@ -22,7 +22,7 @@
 #include <numeric>
 
 using namespace spv;
-using namespace spirv_cross;
+using namespace SPIRV_CROSS_NAMESPACE;
 using namespace std;
 
 static const uint32_t k_unknown_location = ~0u;
@@ -30,7 +30,7 @@ static const uint32_t k_unknown_component = ~0u;
 
 static const uint32_t k_aux_mbr_idx_swizzle_const = 0u;
 
-CompilerMSL::CompilerMSL(vector<uint32_t> spirv_)
+CompilerMSL::CompilerMSL(std::vector<uint32_t> spirv_)
     : CompilerGLSL(move(spirv_))
 {
 }
@@ -423,7 +423,7 @@ void CompilerMSL::emit_entry_point_declarations()
 		if (type.basetype == SPIRType::Sampler)
 			add_resource_name(samp.first);
 
-		vector<string> args;
+		SmallVector<string> args;
 		auto &s = samp.second;
 
 		if (s.coord != MSL_SAMPLER_COORD_NORMALIZED)
@@ -577,13 +577,15 @@ string CompilerMSL::compile()
 	backend.use_initializer_list = true;
 	backend.use_typed_initializer_list = true;
 	backend.native_row_major_matrix = false;
-	backend.flexible_member_array_supported = false;
+	backend.unsized_array_supported = false;
 	backend.can_declare_arrays_inline = false;
 	backend.can_return_array = false;
 	backend.boolean_mix_support = false;
 	backend.allow_truncated_access_chain = true;
 	backend.array_is_value_type = false;
 	backend.comparison_image_samples_scalar = true;
+	backend.native_pointers = true;
+	backend.nonuniform_qualifier = "";
 
 	capture_output_to_buffer = msl_options.capture_output_to_buffer;
 	is_rasterization_disabled = msl_options.disable_rasterization || capture_output_to_buffer;
@@ -659,7 +661,7 @@ string CompilerMSL::compile()
 		next_metal_resource_index_sampler = 0;
 
 		// Move constructor for this type is broken on GCC 4.9 ...
-		buffer = unique_ptr<ostringstream>(new ostringstream());
+		buffer.reset();
 
 		emit_header();
 		emit_specialization_constants_and_structs();
@@ -668,9 +670,9 @@ string CompilerMSL::compile()
 		emit_function(get<SPIRFunction>(ir.default_entry_point), Bitset());
 
 		pass_count++;
-	} while (force_recompile);
+	} while (is_forcing_recompilation());
 
-	return buffer->str();
+	return buffer.str();
 }
 
 // Register the need to output any custom functions.
@@ -679,8 +681,7 @@ void CompilerMSL::preprocess_op_codes()
 	OpCodePreprocessor preproc(*this);
 	traverse_all_reachable_opcodes(get<SPIRFunction>(ir.default_entry_point), preproc);
 
-	if (preproc.suppress_missing_prototypes)
-		add_pragma_line("#pragma clang diagnostic ignored \"-Wmissing-prototypes\"");
+	suppress_missing_prototypes = preproc.suppress_missing_prototypes;
 
 	if (preproc.uses_atomics)
 	{
@@ -1823,7 +1824,7 @@ void CompilerMSL::fix_up_interface_member_indices(StorageClass storage, uint32_t
 uint32_t CompilerMSL::add_interface_block(StorageClass storage, bool patch)
 {
 	// Accumulate the variables that should appear in the interface struct
-	vector<SPIRVariable *> vars;
+	SmallVector<SPIRVariable *> vars;
 	bool incl_builtins = (storage == StorageClassOutput || is_tessellation_shader());
 
 	ir.for_each_typed_id<SPIRVariable>([&](uint32_t var_id, SPIRVariable &var) {
@@ -2351,10 +2352,13 @@ string CompilerMSL::unpack_expression_type(string expr_str, const SPIRType &type
 // Emits the file header info
 void CompilerMSL::emit_header()
 {
+	// This particular line can be overridden during compilation, so make it a flag and not a pragma line.
+	if (suppress_missing_prototypes)
+		statement("#pragma clang diagnostic ignored \"-Wmissing-prototypes\"");
 	for (auto &pragma : pragma_lines)
 		statement(pragma);
 
-	if (!pragma_lines.empty())
+	if (!pragma_lines.empty() || suppress_missing_prototypes)
 		statement("");
 
 	statement("#include <metal_stdlib>");
@@ -2378,14 +2382,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.
@@ -2952,8 +2956,8 @@ void CompilerMSL::emit_specialization_constants_and_structs()
 				// TODO: This can be expressed as a [[threads_per_threadgroup]] input semantic, but we need to know
 				// the work group size at compile time in SPIR-V, and [[threads_per_threadgroup]] would need to be passed around as a global.
 				// The work group size may be a specialization constant.
-				statement("constant uint3 ", builtin_to_glsl(BuiltInWorkgroupSize, StorageClassWorkgroup), " [[maybe_unused]] = ",
-				          constant_expression(get<SPIRConstant>(workgroup_size_id)), ";");
+				statement("constant uint3 ", builtin_to_glsl(BuiltInWorkgroupSize, StorageClassWorkgroup),
+				          " [[maybe_unused]] = ", constant_expression(get<SPIRConstant>(workgroup_size_id)), ";");
 				emitted = true;
 			}
 			else if (c.specialization)
@@ -3080,7 +3084,7 @@ bool CompilerMSL::emit_tessellation_access_chain(const uint32_t *ops, uint32_t l
 	     get_variable_data_type(*var).basetype == SPIRType::Struct))
 	{
 		AccessChainMeta meta;
-		std::vector<uint32_t> indices;
+		SmallVector<uint32_t> indices;
 		uint32_t next_id = ir.increase_bound_by(2);
 
 		indices.reserve(length - 3 + 1);
@@ -3569,7 +3573,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 +3599,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;
@@ -3727,6 +3731,9 @@ void CompilerMSL::emit_instruction(const Instruction &instruction)
 		break;
 	}
 
+	case OpImageTexelPointer:
+		SPIRV_CROSS_THROW("MSL does not support atomic operations on images or texel buffers.");
+
 	// Casting
 	case OpQuantizeToF16:
 	{
@@ -3961,6 +3968,29 @@ void CompilerMSL::emit_array_copy(const string &lhs, uint32_t rhs_id)
 		is_constant = true;
 	}
 
+	// For the case where we have OpLoad triggering an array copy,
+	// we cannot easily detect this case ahead of time since it's
+	// context dependent. We might have to force a recompile here
+	// if this is the only use of array copies in our shader.
+	if (type.array.size() > 1)
+	{
+		if (type.array.size() > SPVFuncImplArrayCopyMultidimMax)
+			SPIRV_CROSS_THROW("Cannot support this many dimensions for arrays of arrays.");
+		auto func = static_cast<SPVFuncImpl>(SPVFuncImplArrayCopyMultidimBase + type.array.size());
+		if (spv_function_implementations.count(func) == 0)
+		{
+			spv_function_implementations.insert(func);
+			suppress_missing_prototypes = true;
+			force_recompile();
+		}
+	}
+	else if (spv_function_implementations.count(SPVFuncImplArrayCopy) == 0)
+	{
+		spv_function_implementations.insert(SPVFuncImplArrayCopy);
+		suppress_missing_prototypes = true;
+		force_recompile();
+	}
+
 	const char *tag = is_constant ? "FromConstant" : "FromStack";
 	statement("spvArrayCopy", tag, type.array.size(), "(", lhs, ", ", to_expression(rhs_id), ");");
 }
@@ -4043,13 +4073,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
@@ -4340,7 +4374,10 @@ void CompilerMSL::emit_function_prototype(SPIRFunction &func, const Bitset &)
 
 		// Manufacture automatic swizzle arg.
 		if (msl_options.swizzle_texture_samples && has_sampled_images && is_sampled_image_type(arg_type))
-			decl += join(", constant uint32_t& ", to_swizzle_expression(arg.id));
+		{
+			bool arg_is_array = !arg_type.array.empty();
+			decl += join(", constant uint32_t", arg_is_array ? "* " : "& ", to_swizzle_expression(arg.id));
+		}
 
 		if (&arg != &func.arguments.back())
 			decl += ", ";
@@ -4454,9 +4491,16 @@ string CompilerMSL::to_function_args(uint32_t img, const SPIRType &imgtype, bool
 		if (coord_type.vecsize > 1)
 			tex_coords = enclose_expression(tex_coords) + ".x";
 
-		// Metal texel buffer textures are 2D, so convert 1D coord to 2D.
-		if (is_fetch)
-			tex_coords = "spvTexelBufferCoord(" + round_fp_tex_coords(tex_coords, coord_is_fp) + ")";
+		if (msl_options.texture_buffer_native)
+		{
+			tex_coords = "uint(" + round_fp_tex_coords(tex_coords, coord_is_fp) + ")";
+		}
+		else
+		{
+			// Metal texel buffer textures are 2D, so convert 1D coord to 2D.
+			if (is_fetch)
+				tex_coords = "spvTexelBufferCoord(" + round_fp_tex_coords(tex_coords, coord_is_fp) + ")";
+		}
 
 		alt_coord_component = 1;
 		break;
@@ -4792,7 +4836,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);
 		}
 	}
@@ -4939,8 +4983,8 @@ void CompilerMSL::add_convert_row_major_matrix_function(uint32_t cols, uint32_t
 	auto rslt = spv_function_implementations.insert(spv_func);
 	if (rslt.second)
 	{
-		add_pragma_line("#pragma clang diagnostic ignored \"-Wmissing-prototypes\"");
-		force_recompile = true;
+		suppress_missing_prototypes = true;
+		force_recompile();
 	}
 }
 
@@ -5579,7 +5623,14 @@ void CompilerMSL::entry_point_args_builtin(string &ep_args)
 			                " [[buffer(", msl_options.shader_output_buffer_index, ")]]");
 		}
 
-		if (stage_out_var_id || get_execution_model() == ExecutionModelTessellationControl)
+		if (get_execution_model() == ExecutionModelTessellationControl)
+		{
+			if (!ep_args.empty())
+				ep_args += ", ";
+			ep_args +=
+			    join("constant uint* spvIndirectParams [[buffer(", msl_options.indirect_params_buffer_index, ")]]");
+		}
+		else if (stage_out_var_id)
 		{
 			if (!ep_args.empty())
 				ep_args += ", ";
@@ -5663,7 +5714,7 @@ void CompilerMSL::entry_point_args_discrete_descriptors(string &ep_args)
 		uint32_t index;
 	};
 
-	vector<Resource> resources;
+	SmallVector<Resource> resources;
 
 	ir.for_each_typed_id<SPIRVariable>([&](uint32_t, SPIRVariable &var) {
 		if ((var.storage == StorageClassUniform || var.storage == StorageClassUniformConstant ||
@@ -5800,9 +5851,13 @@ void CompilerMSL::fix_up_shader_inputs_outputs()
 			if (msl_options.swizzle_texture_samples && has_sampled_images && is_sampled_image_type(type))
 			{
 				auto &entry_func = this->get<SPIRFunction>(ir.default_entry_point);
-				entry_func.fixup_hooks_in.push_back([this, &var, var_id]() {
+				entry_func.fixup_hooks_in.push_back([this, &type, &var, var_id]() {
 					auto &aux_type = expression_type(aux_buffer_id);
-					statement("constant uint32_t& ", to_swizzle_expression(var_id), " = ", to_name(aux_buffer_id), ".",
+					bool is_array_type = !type.array.empty();
+
+					// If we have an array of images, we need to be able to index into it, so take a pointer instead.
+					statement("constant uint32_t", is_array_type ? "* " : "& ", to_swizzle_expression(var_id),
+					          is_array_type ? " = &" : " = ", to_name(aux_buffer_id), ".",
 					          to_member_name(aux_type, k_aux_mbr_idx_swizzle_const), "[",
 					          convert_to_string(get_metal_resource_index(var, SPIRType::Image)), "];");
 				});
@@ -6607,6 +6662,18 @@ string CompilerMSL::image_type_glsl(const SPIRType &type, uint32_t id)
 			img_type_name += (img_type.arrayed ? "texture1d_array" : "texture1d");
 			break;
 		case DimBuffer:
+			if (img_type.ms || img_type.arrayed)
+				SPIRV_CROSS_THROW("Cannot use texel buffers with multisampling or array layers.");
+
+			if (msl_options.texture_buffer_native)
+			{
+				if (!msl_options.supports_msl_version(2, 1))
+					SPIRV_CROSS_THROW("Native texture_buffer type is only supported in MSL 2.1.");
+				img_type_name = "texture_buffer";
+			}
+			else
+				img_type_name += "texture2d";
+			break;
 		case Dim2D:
 		case DimSubpassData:
 			if (img_type.ms && img_type.arrayed)
@@ -7292,7 +7359,7 @@ CompilerMSL::SPVFuncImpl CompilerMSL::OpCodePreprocessor::get_spv_func_impl(Op o
 	{
 		// Retrieve the image type, and if it's a Buffer, emit a texel coordinate function
 		uint32_t tid = result_types[args[opcode == OpImageWrite ? 0 : 2]];
-		if (tid && compiler.get<SPIRType>(tid).image.dim == DimBuffer)
+		if (tid && compiler.get<SPIRType>(tid).image.dim == DimBuffer && !compiler.msl_options.texture_buffer_native)
 			return SPVFuncImplTexelBufferCoords;
 
 		if (opcode == OpImageFetch && compiler.msl_options.swizzle_texture_samples)
@@ -7379,7 +7446,7 @@ void CompilerMSL::MemberSorter::sort()
 	// Create a temporary array of consecutive member indices and sort it based on how
 	// the members should be reordered, based on builtin and sorting aspect meta info.
 	size_t mbr_cnt = type.member_types.size();
-	vector<uint32_t> mbr_idxs(mbr_cnt);
+	SmallVector<uint32_t> mbr_idxs(mbr_cnt);
 	iota(mbr_idxs.begin(), mbr_idxs.end(), 0); // Fill with consecutive indices
 	std::sort(mbr_idxs.begin(), mbr_idxs.end(), *this); // Sort member indices based on sorting aspect
 
@@ -7573,7 +7640,7 @@ void CompilerMSL::analyze_argument_buffers()
 		SPIRType::BaseType basetype;
 		uint32_t index;
 	};
-	vector<Resource> resources_in_set[kMaxArgumentBuffers];
+	SmallVector<Resource> resources_in_set[kMaxArgumentBuffers];
 
 	ir.for_each_typed_id<SPIRVariable>([&](uint32_t self, SPIRVariable &var) {
 		if ((var.storage == StorageClassUniform || var.storage == StorageClassUniformConstant ||
diff --git a/src/3rdparty/SPIRV-Cross/spirv_msl.hpp b/src/3rdparty/SPIRV-Cross/spirv_msl.hpp
index 38610e1..8d3a8ad 100644
--- a/src/3rdparty/SPIRV-Cross/spirv_msl.hpp
+++ b/src/3rdparty/SPIRV-Cross/spirv_msl.hpp
@@ -22,9 +22,8 @@
 #include <set>
 #include <unordered_map>
 #include <unordered_set>
-#include <vector>
 
-namespace spirv_cross
+namespace SPIRV_CROSS_NAMESPACE
 {
 
 // Indicates the format of the vertex attribute. Currently limited to specifying
@@ -195,6 +194,9 @@ public:
 		// Add support to explicit pad out components.
 		bool pad_fragment_output_components = false;
 
+		// Requires MSL 2.1, use the native support for texel buffers.
+		bool texture_buffer_native = false;
+
 		bool is_ios()
 		{
 			return platform == iOS;
@@ -495,9 +497,9 @@ protected:
 	std::unordered_map<MSLStructMemberKey, uint32_t> struct_member_padding;
 	std::set<std::string> pragma_lines;
 	std::set<std::string> typedef_lines;
-	std::vector<uint32_t> vars_needing_early_declaration;
+	SmallVector<uint32_t> vars_needing_early_declaration;
 
-	std::vector<std::pair<MSLResourceBinding, bool>> resource_bindings;
+	SmallVector<std::pair<MSLResourceBinding, bool>> resource_bindings;
 	uint32_t next_metal_resource_index_buffer = 0;
 	uint32_t next_metal_resource_index_texture = 0;
 	uint32_t next_metal_resource_index_sampler = 0;
@@ -530,7 +532,7 @@ protected:
 	spv::Op previous_instruction_opcode = spv::OpNop;
 
 	std::unordered_map<uint32_t, MSLConstexprSampler> constexpr_samplers;
-	std::vector<uint32_t> buffer_arrays;
+	SmallVector<uint32_t> buffer_arrays;
 
 	uint32_t argument_buffer_ids[kMaxArgumentBuffers];
 	uint32_t argument_buffer_discrete_mask = 0;
@@ -540,6 +542,8 @@ protected:
 	uint32_t get_target_components_for_fragment_location(uint32_t location) const;
 	uint32_t build_extended_vector_type(uint32_t type_id, uint32_t components);
 
+	bool suppress_missing_prototypes = false;
+
 	// OpcodeHandler that handles several MSL preprocessing operations.
 	struct OpCodePreprocessor : OpcodeHandler
 	{
@@ -595,6 +599,6 @@ protected:
 		SortAspect sort_aspect;
 	};
 };
-} // namespace spirv_cross
+} // namespace SPIRV_CROSS_NAMESPACE
 
 #endif
diff --git a/src/3rdparty/SPIRV-Cross/spirv_parser.cpp b/src/3rdparty/SPIRV-Cross/spirv_parser.cpp
index fa87fa3..1c0a830 100644
--- a/src/3rdparty/SPIRV-Cross/spirv_parser.cpp
+++ b/src/3rdparty/SPIRV-Cross/spirv_parser.cpp
@@ -20,9 +20,9 @@
 using namespace std;
 using namespace spv;
 
-namespace spirv_cross
+namespace SPIRV_CROSS_NAMESPACE
 {
-Parser::Parser(std::vector<uint32_t> spirv)
+Parser::Parser(vector<uint32_t> spirv)
 {
 	ir.spirv = move(spirv);
 }
@@ -59,6 +59,7 @@ static bool is_valid_spirv_version(uint32_t version)
 	case 0x10100: // SPIR-V 1.1
 	case 0x10200: // SPIR-V 1.2
 	case 0x10300: // SPIR-V 1.3
+	case 0x10400: // SPIR-V 1.4
 		return true;
 
 	default:
@@ -88,7 +89,7 @@ void Parser::parse()
 
 	uint32_t offset = 5;
 
-	vector<Instruction> instructions;
+	SmallVector<Instruction> instructions;
 	while (offset < len)
 	{
 		Instruction instr = {};
@@ -158,7 +159,6 @@ void Parser::parse(const Instruction &instruction)
 
 	switch (op)
 	{
-	case OpMemoryModel:
 	case OpSourceContinued:
 	case OpSourceExtension:
 	case OpNop:
@@ -168,6 +168,11 @@ void Parser::parse(const Instruction &instruction)
 	case OpModuleProcessed:
 		break;
 
+	case OpMemoryModel:
+		ir.addressing_model = static_cast<AddressingModel>(ops[0]);
+		ir.memory_model = static_cast<MemoryModel>(ops[1]);
+		break;
+
 	case OpSource:
 	{
 		auto lang = static_cast<SourceLanguage>(ops[0]);
@@ -207,6 +212,8 @@ void Parser::parse(const Instruction &instruction)
 		uint32_t result_type = ops[0];
 		uint32_t id = ops[1];
 		set<SPIRUndef>(id, result_type);
+		if (current_block)
+			current_block->ops.push_back(instruction);
 		break;
 	}
 
@@ -596,6 +603,20 @@ void Parser::parse(const Instruction &instruction)
 		break;
 	}
 
+	case OpTypeForwardPointer:
+	{
+		uint32_t id = ops[0];
+		auto &ptrbase = set<SPIRType>(id);
+		ptrbase.pointer = true;
+		ptrbase.pointer_depth++;
+		ptrbase.storage = static_cast<StorageClass>(ops[1]);
+
+		if (ptrbase.storage == StorageClassAtomicCounter)
+			ptrbase.basetype = SPIRType::AtomicCounter;
+
+		break;
+	}
+
 	case OpTypeStruct:
 	{
 		uint32_t id = ops[0];
@@ -1095,7 +1116,7 @@ void Parser::make_constant_null(uint32_t id, uint32_t type)
 		if (!constant_type.array_size_literal.back())
 			SPIRV_CROSS_THROW("Array size of OpConstantNull must be a literal.");
 
-		vector<uint32_t> elements(constant_type.array.back());
+		SmallVector<uint32_t> elements(constant_type.array.back());
 		for (uint32_t i = 0; i < constant_type.array.back(); i++)
 			elements[i] = parent_id;
 		set<SPIRConstant>(id, type, elements.data(), uint32_t(elements.size()), false);
@@ -1103,7 +1124,7 @@ void Parser::make_constant_null(uint32_t id, uint32_t type)
 	else if (!constant_type.member_types.empty())
 	{
 		uint32_t member_ids = ir.increase_bound_by(uint32_t(constant_type.member_types.size()));
-		vector<uint32_t> elements(constant_type.member_types.size());
+		SmallVector<uint32_t> elements(constant_type.member_types.size());
 		for (uint32_t i = 0; i < constant_type.member_types.size(); i++)
 		{
 			make_constant_null(member_ids + i, constant_type.member_types[i]);
@@ -1118,4 +1139,4 @@ void Parser::make_constant_null(uint32_t id, uint32_t type)
 	}
 }
 
-} // namespace spirv_cross
+} // namespace SPIRV_CROSS_NAMESPACE
diff --git a/src/3rdparty/SPIRV-Cross/spirv_parser.hpp b/src/3rdparty/SPIRV-Cross/spirv_parser.hpp
index cc15315..ef2c1b9 100644
--- a/src/3rdparty/SPIRV-Cross/spirv_parser.hpp
+++ b/src/3rdparty/SPIRV-Cross/spirv_parser.hpp
@@ -19,9 +19,8 @@
 
 #include "spirv_cross_parsed_ir.hpp"
 #include <stdint.h>
-#include <vector>
 
-namespace spirv_cross
+namespace SPIRV_CROSS_NAMESPACE
 {
 class Parser
 {
@@ -84,12 +83,12 @@ private:
 	}
 
 	// This must be an ordered data structure so we always pick the same type aliases.
-	std::vector<uint32_t> global_struct_cache;
+	SmallVector<uint32_t> global_struct_cache;
 
 	bool types_are_logically_equivalent(const SPIRType &a, const SPIRType &b) const;
 	bool variable_storage_is_aliased(const SPIRVariable &v) const;
 	void make_constant_null(uint32_t id, uint32_t type);
 };
-} // namespace spirv_cross
+} // namespace SPIRV_CROSS_NAMESPACE
 
 #endif
diff --git a/src/3rdparty/SPIRV-Cross/spirv_reflect.cpp b/src/3rdparty/SPIRV-Cross/spirv_reflect.cpp
index c6cd3be..0b2c585 100644
--- a/src/3rdparty/SPIRV-Cross/spirv_reflect.cpp
+++ b/src/3rdparty/SPIRV-Cross/spirv_reflect.cpp
@@ -19,7 +19,7 @@
 #include <iomanip>
 
 using namespace spv;
-using namespace spirv_cross;
+using namespace SPIRV_CROSS_NAMESPACE;
 using namespace std;
 
 namespace simple_json
@@ -36,10 +36,16 @@ using Stack = std::stack<State>;
 class Stream
 {
 	Stack stack;
-	std::ostringstream buffer;
+	StringStream<> buffer;
 	uint32_t indent{ 0 };
+	char current_locale_radix_character = '.';
 
 public:
+	void set_current_locale_radix_character(char c)
+	{
+		current_locale_radix_character = c;
+	}
+
 	void begin_json_object();
 	void end_json_object();
 	void emit_json_key(const std::string &key);
@@ -212,7 +218,7 @@ void Stream::emit_json_key_value(const std::string &key, int32_t value)
 void Stream::emit_json_key_value(const std::string &key, float value)
 {
 	emit_json_key(key);
-	statement_inner(value);
+	statement_inner(convert_to_string(value, current_locale_radix_character));
 }
 
 void Stream::emit_json_key_value(const std::string &key, bool value)
@@ -247,8 +253,8 @@ void CompilerReflection::set_format(const std::string &format)
 
 string CompilerReflection::compile()
 {
-	// Move constructor for this type is broken on GCC 4.9 ...
 	json_stream = std::make_shared<simple_json::Stream>();
+	json_stream->set_current_locale_radix_character(current_locale_radix_character);
 	json_stream->begin_json_object();
 	emit_entry_points();
 	emit_types();
@@ -439,7 +445,7 @@ void CompilerReflection::emit_resources()
 	emit_resources("acceleration_structures", res.acceleration_structures);
 }
 
-void CompilerReflection::emit_resources(const char *tag, const vector<Resource> &resources)
+void CompilerReflection::emit_resources(const char *tag, const SmallVector<Resource> &resources)
 {
 	if (resources.empty())
 	{
diff --git a/src/3rdparty/SPIRV-Cross/spirv_reflect.hpp b/src/3rdparty/SPIRV-Cross/spirv_reflect.hpp
index 13b5b43..5a228a6 100644
--- a/src/3rdparty/SPIRV-Cross/spirv_reflect.hpp
+++ b/src/3rdparty/SPIRV-Cross/spirv_reflect.hpp
@@ -19,14 +19,13 @@
 
 #include "spirv_glsl.hpp"
 #include <utility>
-#include <vector>
 
 namespace simple_json
 {
 class Stream;
 }
 
-namespace spirv_cross
+namespace SPIRV_CROSS_NAMESPACE
 {
 class CompilerReflection : public CompilerGLSL
 {
@@ -34,7 +33,7 @@ class CompilerReflection : public CompilerGLSL
 
 public:
 	explicit CompilerReflection(std::vector<uint32_t> spirv_)
-	    : Parent(move(spirv_))
+	    : Parent(std::move(spirv_))
 	{
 		options.vulkan_semantics = true;
 	}
@@ -72,13 +71,13 @@ private:
 	void emit_type_member(const SPIRType &type, uint32_t index);
 	void emit_type_member_qualifiers(const SPIRType &type, uint32_t index);
 	void emit_type_array(const SPIRType &type);
-	void emit_resources(const char *tag, const std::vector<Resource> &resources);
+	void emit_resources(const char *tag, const SmallVector<Resource> &resources);
 
 	std::string to_member_name(const SPIRType &type, uint32_t index) const;
 
 	std::shared_ptr<simple_json::Stream> json_stream;
 };
 
-} // namespace spirv_cross
+} // namespace SPIRV_CROSS_NAMESPACE
 
 #endif