10 files changed, 398 insertions, 260 deletions

diff --git a/src/shader_recompiler/backend/glsl/emit_context.cpp b/src/shader_recompiler/backend/glsl/emit_context.cpp
index 8eea6344f..b3a3e5647 100644
--- a/src/shader_recompiler/backend/glsl/emit_context.cpp
+++ b/src/shader_recompiler/backend/glsl/emit_context.cpp
@@ -12,6 +12,7 @@ EmitContext::EmitContext(IR::Program& program, [[maybe_unused]] Bindings& bindin
                          const Profile& profile_)
     : info{program.info}, profile{profile_} {
     std::string header = "#version 450\n";
+    SetupExtensions(header);
     if (program.stage == Stage::Compute) {
         header += fmt::format("layout(local_size_x={},local_size_y={},local_size_z={}) in;\n",
                               program.workgroup_size[0], program.workgroup_size[1],
@@ -23,6 +24,12 @@ EmitContext::EmitContext(IR::Program& program, [[maybe_unused]] Bindings& bindin
     code += "void main(){\n";
 }
 
+void EmitContext::SetupExtensions(std::string& header) {
+    if (info.uses_int64) {
+        header += "#extension GL_ARB_gpu_shader_int64 : enable\n";
+    }
+}
+
 void EmitContext::DefineConstantBuffers() {
     if (info.constant_buffer_descriptors.empty()) {
         return;
diff --git a/src/shader_recompiler/backend/glsl/emit_context.h b/src/shader_recompiler/backend/glsl/emit_context.h
index 81b970c14..f8cf8fdbc 100644
--- a/src/shader_recompiler/backend/glsl/emit_context.h
+++ b/src/shader_recompiler/backend/glsl/emit_context.h
@@ -38,28 +38,46 @@ public:
     //     code += '\n';
     // }
 
-    template <typename... Args>
-    void AddU32(const char* format_str, IR::Inst& inst, Args&&... args) {
-        code +=
-            fmt::format(format_str, reg_alloc.Define(inst, Type::U32), std::forward<Args>(args)...);
+    template <Type type, typename... Args>
+    void Add(const char* format_str, IR::Inst& inst, Args&&... args) {
+        code += fmt::format(format_str, reg_alloc.Define(inst, type), std::forward<Args>(args)...);
         // TODO: Remove this
         code += '\n';
     }
 
     template <typename... Args>
+    void AddU1(const char* format_str, IR::Inst& inst, Args&&... args) {
+        Add<Type::U1>(format_str, inst, args...);
+    }
+
+    template <typename... Args>
+    void AddU32(const char* format_str, IR::Inst& inst, Args&&... args) {
+        Add<Type::U32>(format_str, inst, args...);
+    }
+
+    template <typename... Args>
     void AddS32(const char* format_str, IR::Inst& inst, Args&&... args) {
-        code +=
-            fmt::format(format_str, reg_alloc.Define(inst, Type::S32), std::forward<Args>(args)...);
-        // TODO: Remove this
-        code += '\n';
+        Add<Type::S32>(format_str, inst, args...);
     }
 
     template <typename... Args>
     void AddF32(const char* format_str, IR::Inst& inst, Args&&... args) {
-        code +=
-            fmt::format(format_str, reg_alloc.Define(inst, Type::F32), std::forward<Args>(args)...);
-        // TODO: Remove this
-        code += '\n';
+        Add<Type::F32>(format_str, inst, args...);
+    }
+
+    template <typename... Args>
+    void AddU64(const char* format_str, IR::Inst& inst, Args&&... args) {
+        Add<Type::U64>(format_str, inst, args...);
+    }
+
+    template <typename... Args>
+    void AddU32x2(const char* format_str, IR::Inst& inst, Args&&... args) {
+        Add<Type::U32x2>(format_str, inst, args...);
+    }
+
+    template <typename... Args>
+    void AddF32x2(const char* format_str, IR::Inst& inst, Args&&... args) {
+        Add<Type::F32x2>(format_str, inst, args...);
     }
 
     template <typename... Args>
@@ -75,6 +93,7 @@ public:
     const Profile& profile;
 
 private:
+    void SetupExtensions(std::string& header);
     void DefineConstantBuffers();
     void DefineStorageBuffers();
 };
diff --git a/src/shader_recompiler/backend/glsl/emit_glsl_bitwise_conversion.cpp b/src/shader_recompiler/backend/glsl/emit_glsl_bitwise_conversion.cpp
index f40f9900c..73cb66674 100644
--- a/src/shader_recompiler/backend/glsl/emit_glsl_bitwise_conversion.cpp
+++ b/src/shader_recompiler/backend/glsl/emit_glsl_bitwise_conversion.cpp
@@ -25,4 +25,17 @@ static void Alias(IR::Inst& inst, const IR::Value& value) {
 void EmitIdentity(EmitContext&, IR::Inst& inst, const IR::Value& value) {
     Alias(inst, value);
 }
+
+void EmitConditionRef(EmitContext& ctx, IR::Inst& inst, const IR::Value& value) {
+    ctx.AddU1("{}={};", inst, ctx.reg_alloc.Consume(value));
+}
+
+void EmitPackUint2x32(EmitContext& ctx, IR::Inst& inst, std::string_view value) {
+    ctx.AddU64("{}=packUint2x32({});", inst, value);
+}
+
+void EmitUnpackUint2x32(EmitContext& ctx, IR::Inst& inst, std::string_view value) {
+    ctx.AddU32x2("{}=unpackUint2x32({});", inst, value);
+}
+
 } // namespace Shader::Backend::GLSL
diff --git a/src/shader_recompiler/backend/glsl/emit_glsl_composite.cpp b/src/shader_recompiler/backend/glsl/emit_glsl_composite.cpp
index e69de29bb..2a7d207a7 100644
--- a/src/shader_recompiler/backend/glsl/emit_glsl_composite.cpp
+++ b/src/shader_recompiler/backend/glsl/emit_glsl_composite.cpp
@@ -0,0 +1,237 @@
+// Copyright 2021 yuzu Emulator Project
+// Licensed under GPLv2 or any later version
+// Refer to the license.txt file included.
+
+#include <string_view>
+
+#include "shader_recompiler/backend/glsl/emit_context.h"
+#include "shader_recompiler/backend/glsl/emit_glsl_instructions.h"
+#include "shader_recompiler/frontend/ir/value.h"
+#include "shader_recompiler/profile.h"
+
+namespace Shader::Backend::GLSL {
+void EmitCompositeConstructU32x2([[maybe_unused]] EmitContext& ctx, IR::Inst& inst,
+                                 [[maybe_unused]] std::string_view e1,
+                                 [[maybe_unused]] std::string_view e2) {
+    ctx.AddU32x2("{}=uvec2({},{});", inst, e1, e2);
+}
+
+void EmitCompositeConstructU32x3([[maybe_unused]] EmitContext& ctx,
+                                 [[maybe_unused]] std::string_view e1,
+                                 [[maybe_unused]] std::string_view e2,
+                                 [[maybe_unused]] std::string_view e3) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeConstructU32x4([[maybe_unused]] EmitContext& ctx,
+                                 [[maybe_unused]] std::string_view e1,
+                                 [[maybe_unused]] std::string_view e2,
+                                 [[maybe_unused]] std::string_view e3,
+                                 [[maybe_unused]] std::string_view e4) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeExtractU32x2([[maybe_unused]] EmitContext& ctx, IR::Inst& inst,
+                               [[maybe_unused]] std::string_view composite,
+                               [[maybe_unused]] u32 index) {
+    ctx.AddU32("{}={}[{}];", inst, composite, index);
+}
+
+void EmitCompositeExtractU32x3([[maybe_unused]] EmitContext& ctx,
+                               [[maybe_unused]] std::string_view composite,
+                               [[maybe_unused]] u32 index) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeExtractU32x4([[maybe_unused]] EmitContext& ctx,
+                               [[maybe_unused]] std::string_view composite,
+                               [[maybe_unused]] u32 index) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeInsertU32x2([[maybe_unused]] EmitContext& ctx,
+                              [[maybe_unused]] std::string_view composite,
+                              [[maybe_unused]] std::string_view object,
+                              [[maybe_unused]] u32 index) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeInsertU32x3([[maybe_unused]] EmitContext& ctx,
+                              [[maybe_unused]] std::string_view composite,
+                              [[maybe_unused]] std::string_view object,
+                              [[maybe_unused]] u32 index) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeInsertU32x4([[maybe_unused]] EmitContext& ctx,
+                              [[maybe_unused]] std::string_view composite,
+                              [[maybe_unused]] std::string_view object,
+                              [[maybe_unused]] u32 index) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeConstructF16x2([[maybe_unused]] EmitContext& ctx,
+                                 [[maybe_unused]] std::string_view e1,
+                                 [[maybe_unused]] std::string_view e2) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeConstructF16x3([[maybe_unused]] EmitContext& ctx,
+                                 [[maybe_unused]] std::string_view e1,
+                                 [[maybe_unused]] std::string_view e2,
+                                 [[maybe_unused]] std::string_view e3) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeConstructF16x4([[maybe_unused]] EmitContext& ctx,
+                                 [[maybe_unused]] std::string_view e1,
+                                 [[maybe_unused]] std::string_view e2,
+                                 [[maybe_unused]] std::string_view e3,
+                                 [[maybe_unused]] std::string_view e4) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeExtractF16x2([[maybe_unused]] EmitContext& ctx,
+                               [[maybe_unused]] std::string_view composite,
+                               [[maybe_unused]] u32 index) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeExtractF16x3([[maybe_unused]] EmitContext& ctx,
+                               [[maybe_unused]] std::string_view composite,
+                               [[maybe_unused]] u32 index) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeExtractF16x4([[maybe_unused]] EmitContext& ctx,
+                               [[maybe_unused]] std::string_view composite,
+                               [[maybe_unused]] u32 index) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeInsertF16x2([[maybe_unused]] EmitContext& ctx,
+                              [[maybe_unused]] std::string_view composite,
+                              [[maybe_unused]] std::string_view object,
+                              [[maybe_unused]] u32 index) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeInsertF16x3([[maybe_unused]] EmitContext& ctx,
+                              [[maybe_unused]] std::string_view composite,
+                              [[maybe_unused]] std::string_view object,
+                              [[maybe_unused]] u32 index) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeInsertF16x4([[maybe_unused]] EmitContext& ctx,
+                              [[maybe_unused]] std::string_view composite,
+                              [[maybe_unused]] std::string_view object,
+                              [[maybe_unused]] u32 index) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeConstructF32x2([[maybe_unused]] EmitContext& ctx,
+                                 [[maybe_unused]] std::string_view e1,
+                                 [[maybe_unused]] std::string_view e2) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeConstructF32x3([[maybe_unused]] EmitContext& ctx,
+                                 [[maybe_unused]] std::string_view e1,
+                                 [[maybe_unused]] std::string_view e2,
+                                 [[maybe_unused]] std::string_view e3) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeConstructF32x4([[maybe_unused]] EmitContext& ctx,
+                                 [[maybe_unused]] std::string_view e1,
+                                 [[maybe_unused]] std::string_view e2,
+                                 [[maybe_unused]] std::string_view e3,
+                                 [[maybe_unused]] std::string_view e4) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeExtractF32x2([[maybe_unused]] EmitContext& ctx,
+                               [[maybe_unused]] std::string_view composite,
+                               [[maybe_unused]] u32 index) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeExtractF32x3([[maybe_unused]] EmitContext& ctx,
+                               [[maybe_unused]] std::string_view composite,
+                               [[maybe_unused]] u32 index) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeExtractF32x4([[maybe_unused]] EmitContext& ctx,
+                               [[maybe_unused]] std::string_view composite,
+                               [[maybe_unused]] u32 index) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeInsertF32x2([[maybe_unused]] EmitContext& ctx,
+                              [[maybe_unused]] std::string_view composite,
+                              [[maybe_unused]] std::string_view object,
+                              [[maybe_unused]] u32 index) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeInsertF32x3([[maybe_unused]] EmitContext& ctx,
+                              [[maybe_unused]] std::string_view composite,
+                              [[maybe_unused]] std::string_view object,
+                              [[maybe_unused]] u32 index) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeInsertF32x4([[maybe_unused]] EmitContext& ctx,
+                              [[maybe_unused]] std::string_view composite,
+                              [[maybe_unused]] std::string_view object,
+                              [[maybe_unused]] u32 index) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeConstructF64x2([[maybe_unused]] EmitContext& ctx) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeConstructF64x3([[maybe_unused]] EmitContext& ctx) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeConstructF64x4([[maybe_unused]] EmitContext& ctx) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeExtractF64x2([[maybe_unused]] EmitContext& ctx) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeExtractF64x3([[maybe_unused]] EmitContext& ctx) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeExtractF64x4([[maybe_unused]] EmitContext& ctx) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeInsertF64x2([[maybe_unused]] EmitContext& ctx,
+                              [[maybe_unused]] std::string_view composite,
+                              [[maybe_unused]] std::string_view object,
+                              [[maybe_unused]] u32 index) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeInsertF64x3([[maybe_unused]] EmitContext& ctx,
+                              [[maybe_unused]] std::string_view composite,
+                              [[maybe_unused]] std::string_view object,
+                              [[maybe_unused]] u32 index) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitCompositeInsertF64x4([[maybe_unused]] EmitContext& ctx,
+                              [[maybe_unused]] std::string_view composite,
+                              [[maybe_unused]] std::string_view object,
+                              [[maybe_unused]] u32 index) {
+    throw NotImplementedException("GLSL Instruction");
+}
+} // namespace Shader::Backend::GLSL
diff --git a/src/shader_recompiler/backend/glsl/emit_glsl_instructions.h b/src/shader_recompiler/backend/glsl/emit_glsl_instructions.h
index ff0c9cd95..51dbeb2c1 100644
--- a/src/shader_recompiler/backend/glsl/emit_glsl_instructions.h
+++ b/src/shader_recompiler/backend/glsl/emit_glsl_instructions.h
@@ -142,12 +142,14 @@ void EmitWriteSharedU16(EmitContext& ctx, std::string_view offset, std::string_v
 void EmitWriteSharedU32(EmitContext& ctx, std::string_view offset, std::string_view value);
 void EmitWriteSharedU64(EmitContext& ctx, std::string_view offset, std::string_view value);
 void EmitWriteSharedU128(EmitContext& ctx, std::string_view offset, std::string_view value);
-void EmitCompositeConstructU32x2(EmitContext& ctx, std::string_view e1, std::string_view e2);
+void EmitCompositeConstructU32x2(EmitContext& ctx, IR::Inst& inst, std::string_view e1,
+                                 std::string_view e2);
 void EmitCompositeConstructU32x3(EmitContext& ctx, std::string_view e1, std::string_view e2,
                                  std::string_view e3);
 void EmitCompositeConstructU32x4(EmitContext& ctx, std::string_view e1, std::string_view e2,
                                  std::string_view e3, std::string_view e4);
-void EmitCompositeExtractU32x2(EmitContext& ctx, std::string_view composite, u32 index);
+void EmitCompositeExtractU32x2(EmitContext& ctx, IR::Inst& inst, std::string_view composite,
+                               u32 index);
 void EmitCompositeExtractU32x3(EmitContext& ctx, std::string_view composite, u32 index);
 void EmitCompositeExtractU32x4(EmitContext& ctx, std::string_view composite, u32 index);
 void EmitCompositeInsertU32x2(EmitContext& ctx, std::string_view composite, std::string_view object,
@@ -218,8 +220,8 @@ void EmitBitCastU64F64(EmitContext& ctx);
 void EmitBitCastF16U16(EmitContext& ctx);
 void EmitBitCastF32U32(EmitContext& ctx, std::string_view value);
 void EmitBitCastF64U64(EmitContext& ctx);
-void EmitPackUint2x32(EmitContext& ctx, std::string_view value);
-void EmitUnpackUint2x32(EmitContext& ctx, std::string_view value);
+void EmitPackUint2x32(EmitContext& ctx, IR::Inst& inst, std::string_view value);
+void EmitUnpackUint2x32(EmitContext& ctx, IR::Inst& inst, std::string_view value);
 void EmitPackFloat2x16(EmitContext& ctx, std::string_view value);
 void EmitUnpackFloat2x16(EmitContext& ctx, std::string_view value);
 void EmitPackHalf2x16(EmitContext& ctx, std::string_view value);
diff --git a/src/shader_recompiler/backend/glsl/emit_glsl_integer.cpp b/src/shader_recompiler/backend/glsl/emit_glsl_integer.cpp
index a22313141..016bccd39 100644
--- a/src/shader_recompiler/backend/glsl/emit_glsl_integer.cpp
+++ b/src/shader_recompiler/backend/glsl/emit_glsl_integer.cpp
@@ -1,4 +1,3 @@
-
 // Copyright 2021 yuzu Emulator Project
 // Licensed under GPLv2 or any later version
 // Refer to the license.txt file included.
@@ -48,7 +47,7 @@ void EmitINeg64([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst& in
 
 void EmitIAbs32([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst& inst,
                 [[maybe_unused]] std::string_view value) {
-    throw NotImplementedException("GLSL Instruction");
+    ctx.AddU32("{}=abs({});", inst, value);
 }
 
 void EmitIAbs64([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst& inst,
@@ -59,52 +58,52 @@ void EmitIAbs64([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst& in
 void EmitShiftLeftLogical32([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst& inst,
                             [[maybe_unused]] std::string_view base,
                             [[maybe_unused]] std::string_view shift) {
-    throw NotImplementedException("GLSL Instruction");
+    ctx.AddU32("{}={}<<{};", inst, base, shift);
 }
 
 void EmitShiftLeftLogical64([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst& inst,
                             [[maybe_unused]] std::string_view base,
                             [[maybe_unused]] std::string_view shift) {
-    throw NotImplementedException("GLSL Instruction");
+    ctx.AddU64("{}={}<<{};", inst, base, shift);
 }
 
 void EmitShiftRightLogical32([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst& inst,
                              [[maybe_unused]] std::string_view base,
                              [[maybe_unused]] std::string_view shift) {
-    throw NotImplementedException("GLSL Instruction");
+    ctx.AddU32("{}={}>>{};", inst, base, shift);
 }
 
 void EmitShiftRightLogical64([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst& inst,
                              [[maybe_unused]] std::string_view base,
                              [[maybe_unused]] std::string_view shift) {
-    throw NotImplementedException("GLSL Instruction");
+    ctx.AddU64("{}={}>>{};", inst, base, shift);
 }
 
 void EmitShiftRightArithmetic32([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst& inst,
                                 [[maybe_unused]] std::string_view base,
                                 [[maybe_unused]] std::string_view shift) {
-    throw NotImplementedException("GLSL Instruction");
+    ctx.AddS32("{}=int({})>>{};", inst, base, shift);
 }
 
 void EmitShiftRightArithmetic64([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst& inst,
                                 [[maybe_unused]] std::string_view base,
                                 [[maybe_unused]] std::string_view shift) {
-    throw NotImplementedException("GLSL Instruction");
+    ctx.AddU64("{}=int64_t({})>>{};", inst, base, shift);
 }
 
 void EmitBitwiseAnd32([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst& inst,
                       [[maybe_unused]] std::string_view a, [[maybe_unused]] std::string_view b) {
-    throw NotImplementedException("GLSL Instruction");
+    ctx.AddU32("{}={}&{};", inst, a, b);
 }
 
 void EmitBitwiseOr32([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst& inst,
                      [[maybe_unused]] std::string_view a, [[maybe_unused]] std::string_view b) {
-    throw NotImplementedException("GLSL Instruction");
+    ctx.AddU32("{}={}|{};", inst, a, b);
 }
 
 void EmitBitwiseXor32([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst& inst,
                       [[maybe_unused]] std::string_view a, [[maybe_unused]] std::string_view b) {
-    throw NotImplementedException("GLSL Instruction");
+    ctx.AddU32("{}={}^{};", inst, a, b);
 }
 
 void EmitBitFieldInsert([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst& inst,
@@ -141,7 +140,7 @@ void EmitBitCount32([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst
 
 void EmitBitwiseNot32([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst& inst,
                       [[maybe_unused]] std::string_view value) {
-    throw NotImplementedException("GLSL Instruction");
+    ctx.AddU32("{}=~{};", inst, value);
 }
 
 void EmitFindSMsb32([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst& inst,
@@ -156,22 +155,22 @@ void EmitFindUMsb32([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst
 
 void EmitSMin32([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst& inst,
                 [[maybe_unused]] std::string_view a, [[maybe_unused]] std::string_view b) {
-    throw NotImplementedException("GLSL Instruction");
+    ctx.AddU32("{}=min(int({}), int({}));", inst, a, b);
 }
 
 void EmitUMin32([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst& inst,
                 [[maybe_unused]] std::string_view a, [[maybe_unused]] std::string_view b) {
-    throw NotImplementedException("GLSL Instruction");
+    ctx.AddU32("{}=min(uint({}), uint({}));", inst, a, b);
 }
 
 void EmitSMax32([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst& inst,
                 [[maybe_unused]] std::string_view a, [[maybe_unused]] std::string_view b) {
-    throw NotImplementedException("GLSL Instruction");
+    ctx.AddU32("{}=max(int({}), int({}));", inst, a, b);
 }
 
 void EmitUMax32([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst& inst,
                 [[maybe_unused]] std::string_view a, [[maybe_unused]] std::string_view b) {
-    throw NotImplementedException("GLSL Instruction");
+    ctx.AddU32("{}=max(uint({}), uint({}));", inst, a, b);
 }
 
 void EmitSClamp32([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst& inst,
@@ -188,57 +187,57 @@ void EmitUClamp32([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst&
 
 void EmitSLessThan([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst& inst,
                    [[maybe_unused]] std::string_view lhs, [[maybe_unused]] std::string_view rhs) {
-    throw NotImplementedException("GLSL Instruction");
+    ctx.AddU1("{}=int({})<int({});", inst, lhs, rhs);
 }
 
 void EmitULessThan([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst& inst,
                    [[maybe_unused]] std::string_view lhs, [[maybe_unused]] std::string_view rhs) {
-    throw NotImplementedException("GLSL Instruction");
+    ctx.AddU1("{}=uint({})<uint({)};", inst, lhs, rhs);
 }
 
 void EmitIEqual([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst& inst,
                 [[maybe_unused]] std::string_view lhs, [[maybe_unused]] std::string_view rhs) {
-    throw NotImplementedException("GLSL Instruction");
+    ctx.AddU1("{}={}=={};", inst, lhs, rhs);
 }
 
 void EmitSLessThanEqual([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst& inst,
                         [[maybe_unused]] std::string_view lhs,
                         [[maybe_unused]] std::string_view rhs) {
-    throw NotImplementedException("GLSL Instruction");
+    ctx.AddU1("{}=int({})<=int({});", inst, lhs, rhs);
 }
 
 void EmitULessThanEqual([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst& inst,
                         [[maybe_unused]] std::string_view lhs,
                         [[maybe_unused]] std::string_view rhs) {
-    throw NotImplementedException("GLSL Instruction");
+    ctx.AddU1("{}=uint({})<=uint({});", inst, lhs, rhs);
 }
 
 void EmitSGreaterThan([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst& inst,
                       [[maybe_unused]] std::string_view lhs,
                       [[maybe_unused]] std::string_view rhs) {
-    throw NotImplementedException("GLSL Instruction");
+    ctx.AddU1("{}=int({})>int({});", inst, lhs, rhs);
 }
 
 void EmitUGreaterThan([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst& inst,
                       [[maybe_unused]] std::string_view lhs,
                       [[maybe_unused]] std::string_view rhs) {
-    throw NotImplementedException("GLSL Instruction");
+    ctx.AddU1("{}=uint({})>uint({});", inst, lhs, rhs);
 }
 
 void EmitINotEqual([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst& inst,
                    [[maybe_unused]] std::string_view lhs, [[maybe_unused]] std::string_view rhs) {
-    throw NotImplementedException("GLSL Instruction");
+    ctx.AddU1("{}={}!={};", inst, lhs, rhs);
 }
 
 void EmitSGreaterThanEqual([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst& inst,
                            [[maybe_unused]] std::string_view lhs,
                            [[maybe_unused]] std::string_view rhs) {
-    throw NotImplementedException("GLSL Instruction");
+    ctx.AddU1("{}=int({})>=int({});", inst, lhs, rhs);
 }
 
 void EmitUGreaterThanEqual([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] IR::Inst& inst,
                            [[maybe_unused]] std::string_view lhs,
                            [[maybe_unused]] std::string_view rhs) {
-    throw NotImplementedException("GLSL Instruction");
+    ctx.AddU1("{}=uint({})>=uint({});", inst, lhs, rhs);
 }
 } // namespace Shader::Backend::GLSL
diff --git a/src/shader_recompiler/backend/glsl/emit_glsl_not_implemented.cpp b/src/shader_recompiler/backend/glsl/emit_glsl_not_implemented.cpp
index ff2b30eee..de350b154 100644
--- a/src/shader_recompiler/backend/glsl/emit_glsl_not_implemented.cpp
+++ b/src/shader_recompiler/backend/glsl/emit_glsl_not_implemented.cpp
@@ -27,10 +27,6 @@ void EmitVoid(EmitContext& ctx) {
     NotImplemented();
 }
 
-void EmitConditionRef(EmitContext& ctx, IR::Inst& inst, const IR::Value& value) {
-    NotImplemented();
-}
-
 void EmitReference(EmitContext&) {
     NotImplemented();
 }
@@ -359,208 +355,6 @@ void EmitWriteSharedU128(EmitContext& ctx, std::string_view offset, std::string_
     NotImplemented();
 }
 
-void EmitCompositeConstructU32x2(EmitContext& ctx, std::string_view e1, std::string_view e2) {
-    NotImplemented();
-}
-
-void EmitCompositeConstructU32x3(EmitContext& ctx, std::string_view e1, std::string_view e2,
-                                 std::string_view e3) {
-    NotImplemented();
-}
-
-void EmitCompositeConstructU32x4(EmitContext& ctx, std::string_view e1, std::string_view e2,
-                                 std::string_view e3, std::string_view e4) {
-    NotImplemented();
-}
-
-void EmitCompositeExtractU32x2(EmitContext& ctx, std::string_view composite, u32 index) {
-    NotImplemented();
-}
-
-void EmitCompositeExtractU32x3(EmitContext& ctx, std::string_view composite, u32 index) {
-    NotImplemented();
-}
-
-void EmitCompositeExtractU32x4(EmitContext& ctx, std::string_view composite, u32 index) {
-    NotImplemented();
-}
-
-void EmitCompositeInsertU32x2(EmitContext& ctx, std::string_view composite, std::string_view object,
-                              u32 index) {
-    NotImplemented();
-}
-
-void EmitCompositeInsertU32x3(EmitContext& ctx, std::string_view composite, std::string_view object,
-                              u32 index) {
-    NotImplemented();
-}
-
-void EmitCompositeInsertU32x4(EmitContext& ctx, std::string_view composite, std::string_view object,
-                              u32 index) {
-    NotImplemented();
-}
-
-void EmitCompositeConstructF16x2(EmitContext& ctx, std::string_view e1, std::string_view e2) {
-    NotImplemented();
-}
-
-void EmitCompositeConstructF16x3(EmitContext& ctx, std::string_view e1, std::string_view e2,
-                                 std::string_view e3) {
-    NotImplemented();
-}
-
-void EmitCompositeConstructF16x4(EmitContext& ctx, std::string_view e1, std::string_view e2,
-                                 std::string_view e3, std::string_view e4) {
-    NotImplemented();
-}
-
-void EmitCompositeExtractF16x2(EmitContext& ctx, std::string_view composite, u32 index) {
-    NotImplemented();
-}
-
-void EmitCompositeExtractF16x3(EmitContext& ctx, std::string_view composite, u32 index) {
-    NotImplemented();
-}
-
-void EmitCompositeExtractF16x4(EmitContext& ctx, std::string_view composite, u32 index) {
-    NotImplemented();
-}
-
-void EmitCompositeInsertF16x2(EmitContext& ctx, std::string_view composite, std::string_view object,
-                              u32 index) {
-    NotImplemented();
-}
-
-void EmitCompositeInsertF16x3(EmitContext& ctx, std::string_view composite, std::string_view object,
-                              u32 index) {
-    NotImplemented();
-}
-
-void EmitCompositeInsertF16x4(EmitContext& ctx, std::string_view composite, std::string_view object,
-                              u32 index) {
-    NotImplemented();
-}
-
-void EmitCompositeConstructF32x2(EmitContext& ctx, std::string_view e1, std::string_view e2) {
-    NotImplemented();
-}
-
-void EmitCompositeConstructF32x3(EmitContext& ctx, std::string_view e1, std::string_view e2,
-                                 std::string_view e3) {
-    NotImplemented();
-}
-
-void EmitCompositeConstructF32x4(EmitContext& ctx, std::string_view e1, std::string_view e2,
-                                 std::string_view e3, std::string_view e4) {
-    NotImplemented();
-}
-
-void EmitCompositeExtractF32x2(EmitContext& ctx, std::string_view composite, u32 index) {
-    NotImplemented();
-}
-
-void EmitCompositeExtractF32x3(EmitContext& ctx, std::string_view composite, u32 index) {
-    NotImplemented();
-}
-
-void EmitCompositeExtractF32x4(EmitContext& ctx, std::string_view composite, u32 index) {
-    NotImplemented();
-}
-
-void EmitCompositeInsertF32x2(EmitContext& ctx, std::string_view composite, std::string_view object,
-                              u32 index) {
-    NotImplemented();
-}
-
-void EmitCompositeInsertF32x3(EmitContext& ctx, std::string_view composite, std::string_view object,
-                              u32 index) {
-    NotImplemented();
-}
-
-void EmitCompositeInsertF32x4(EmitContext& ctx, std::string_view composite, std::string_view object,
-                              u32 index) {
-    NotImplemented();
-}
-
-void EmitCompositeConstructF64x2(EmitContext& ctx) {
-    NotImplemented();
-}
-
-void EmitCompositeConstructF64x3(EmitContext& ctx) {
-    NotImplemented();
-}
-
-void EmitCompositeConstructF64x4(EmitContext& ctx) {
-    NotImplemented();
-}
-
-void EmitCompositeExtractF64x2(EmitContext& ctx) {
-    NotImplemented();
-}
-
-void EmitCompositeExtractF64x3(EmitContext& ctx) {
-    NotImplemented();
-}
-
-void EmitCompositeExtractF64x4(EmitContext& ctx) {
-    NotImplemented();
-}
-
-void EmitCompositeInsertF64x2(EmitContext& ctx, std::string_view composite, std::string_view object,
-                              u32 index) {
-    NotImplemented();
-}
-
-void EmitCompositeInsertF64x3(EmitContext& ctx, std::string_view composite, std::string_view object,
-                              u32 index) {
-    NotImplemented();
-}
-
-void EmitCompositeInsertF64x4(EmitContext& ctx, std::string_view composite, std::string_view object,
-                              u32 index) {
-    NotImplemented();
-}
-
-void EmitSelectU1(EmitContext& ctx, std::string_view cond, std::string_view true_value,
-                  std::string_view false_value) {
-    NotImplemented();
-}
-
-void EmitSelectU8(EmitContext& ctx, std::string_view cond, std::string_view true_value,
-                  std::string_view false_value) {
-    NotImplemented();
-}
-
-void EmitSelectU16(EmitContext& ctx, std::string_view cond, std::string_view true_value,
-                   std::string_view false_value) {
-    NotImplemented();
-}
-
-void EmitSelectU32(EmitContext& ctx, std::string_view cond, std::string_view true_value,
-                   std::string_view false_value) {
-    NotImplemented();
-}
-
-void EmitSelectU64(EmitContext& ctx, std::string_view cond, std::string_view true_value,
-                   std::string_view false_value) {
-    NotImplemented();
-}
-
-void EmitSelectF16(EmitContext& ctx, std::string_view cond, std::string_view true_value,
-                   std::string_view false_value) {
-    NotImplemented();
-}
-
-void EmitSelectF32(EmitContext& ctx, std::string_view cond, std::string_view true_value,
-                   std::string_view false_value) {
-    NotImplemented();
-}
-
-void EmitSelectF64(EmitContext& ctx, std::string_view cond, std::string_view true_value,
-                   std::string_view false_value) {
-    NotImplemented();
-}
-
 void EmitBitCastU16F16(EmitContext& ctx) {
     NotImplemented();
 }
@@ -585,14 +379,6 @@ void EmitBitCastF64U64(EmitContext& ctx) {
     NotImplemented();
 }
 
-void EmitPackUint2x32(EmitContext& ctx, std::string_view value) {
-    NotImplemented();
-}
-
-void EmitUnpackUint2x32(EmitContext& ctx, std::string_view value) {
-    NotImplemented();
-}
-
 void EmitPackFloat2x16(EmitContext& ctx, std::string_view value) {
     NotImplemented();
 }
diff --git a/src/shader_recompiler/backend/glsl/emit_glsl_select.cpp b/src/shader_recompiler/backend/glsl/emit_glsl_select.cpp
index e69de29bb..86d38da98 100644
--- a/src/shader_recompiler/backend/glsl/emit_glsl_select.cpp
+++ b/src/shader_recompiler/backend/glsl/emit_glsl_select.cpp
@@ -0,0 +1,61 @@
+// Copyright 2021 yuzu Emulator Project
+// Licensed under GPLv2 or any later version
+// Refer to the license.txt file included.
+
+#include <string_view>
+
+#include "shader_recompiler/backend/glsl/emit_context.h"
+#include "shader_recompiler/backend/glsl/emit_glsl_instructions.h"
+#include "shader_recompiler/frontend/ir/value.h"
+#include "shader_recompiler/profile.h"
+
+namespace Shader::Backend::GLSL {
+void EmitSelectU1([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] std::string_view cond,
+                  [[maybe_unused]] std::string_view true_value,
+                  [[maybe_unused]] std::string_view false_value) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitSelectU8([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] std::string_view cond,
+                  [[maybe_unused]] std::string_view true_value,
+                  [[maybe_unused]] std::string_view false_value) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitSelectU16([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] std::string_view cond,
+                   [[maybe_unused]] std::string_view true_value,
+                   [[maybe_unused]] std::string_view false_value) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitSelectU32([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] std::string_view cond,
+                   [[maybe_unused]] std::string_view true_value,
+                   [[maybe_unused]] std::string_view false_value) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitSelectU64([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] std::string_view cond,
+                   [[maybe_unused]] std::string_view true_value,
+                   [[maybe_unused]] std::string_view false_value) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitSelectF16([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] std::string_view cond,
+                   [[maybe_unused]] std::string_view true_value,
+                   [[maybe_unused]] std::string_view false_value) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitSelectF32([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] std::string_view cond,
+                   [[maybe_unused]] std::string_view true_value,
+                   [[maybe_unused]] std::string_view false_value) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+void EmitSelectF64([[maybe_unused]] EmitContext& ctx, [[maybe_unused]] std::string_view cond,
+                   [[maybe_unused]] std::string_view true_value,
+                   [[maybe_unused]] std::string_view false_value) {
+    throw NotImplementedException("GLSL Instruction");
+}
+
+} // namespace Shader::Backend::GLSL
diff --git a/src/shader_recompiler/backend/glsl/reg_alloc.cpp b/src/shader_recompiler/backend/glsl/reg_alloc.cpp
index f4886dbfd..3cfa16fea 100644
--- a/src/shader_recompiler/backend/glsl/reg_alloc.cpp
+++ b/src/shader_recompiler/backend/glsl/reg_alloc.cpp
@@ -57,9 +57,10 @@ std::string RegAlloc::Consume(const IR::Value& value) {
 std::string RegAlloc::Consume(IR::Inst& inst) {
     const Id id{inst.Definition<Id>()};
     inst.DestructiveRemoveUsage();
-    if (!inst.HasUses()) {
-        Free(id);
-    }
+    // TODO: reuse variables of same type if possible
+    // if (!inst.HasUses()) {
+    //     Free(id);
+    // }
     return Representation(inst.Definition<Id>());
 }
 
@@ -69,14 +70,24 @@ std::string RegAlloc::GetType(Type type, u32 index) {
     }
     register_defined[index] = true;
     switch (type) {
+    case Type::U1:
+        return "bool ";
     case Type::U32:
         return "uint ";
     case Type::S32:
         return "int ";
     case Type::F32:
         return "float ";
-    default:
+    case Type::U64:
+        return "uint64_t ";
+    case Type::U32x2:
+        return "uvec2 ";
+    case Type::F32x2:
+        return "vec2 ";
+    case Type::Void:
         return "";
+    default:
+        throw NotImplementedException("Type {}", type);
     }
 }
 
diff --git a/src/shader_recompiler/backend/glsl/reg_alloc.h b/src/shader_recompiler/backend/glsl/reg_alloc.h
index 9b98aab39..84ef7e822 100644
--- a/src/shader_recompiler/backend/glsl/reg_alloc.h
+++ b/src/shader_recompiler/backend/glsl/reg_alloc.h
@@ -16,11 +16,14 @@ class Value;
 
 namespace Shader::Backend::GLSL {
 enum class Type : u32 {
+    U1,
     U32,
     S32,
     F32,
     U64,
     F64,
+    U32x2,
+    F32x2,
     Void,
 };