ShaderCompiler: Inline driver specific constants.

5 files changed, 39 insertions, 3 deletions

diff --git a/src/shader_recompiler/environment.h b/src/shader_recompiler/environment.h
index 8fc359126..26e8307c1 100644
--- a/src/shader_recompiler/environment.h
+++ b/src/shader_recompiler/environment.h
@@ -57,11 +57,16 @@ public:
         return start_address;
     }
 
+    [[nodiscard]] bool IsPropietaryDriver() const noexcept {
+        return is_propietary_driver;
+    }
+
 protected:
     ProgramHeader sph{};
     std::array<u32, 8> gp_passthrough_mask{};
     Stage stage{};
     u32 start_address{};
+    bool is_propietary_driver{};
 };
 
 } // namespace Shader
diff --git a/src/shader_recompiler/ir_opt/constant_propagation_pass.cpp b/src/shader_recompiler/ir_opt/constant_propagation_pass.cpp
index ac10405f3..5275b2c8b 100644
--- a/src/shader_recompiler/ir_opt/constant_propagation_pass.cpp
+++ b/src/shader_recompiler/ir_opt/constant_propagation_pass.cpp
@@ -677,6 +677,30 @@ void FoldConstBuffer(Environment& env, IR::Block& block, IR::Inst& inst) {
     }
 }
 
+void FoldDriverConstBuffer(Environment& env, IR::Block& block, IR::Inst& inst, u32 which_bank,
+                           u32 offset_start = 0, u32 offset_end = std::numeric_limits<u16>::max()) {
+    const IR::Value bank{inst.Arg(0)};
+    const IR::Value offset{inst.Arg(1)};
+    if (!bank.IsImmediate() || !offset.IsImmediate()) {
+        return;
+    }
+    const auto bank_value = bank.U32();
+    if (bank_value != which_bank) {
+        return;
+    }
+    const auto offset_value = offset.U32();
+    if (offset_value < offset_start || offset_value >= offset_end) {
+        return;
+    }
+    IR::IREmitter ir{block, IR::Block::InstructionList::s_iterator_to(inst)};
+    if (inst.GetOpcode() == IR::Opcode::GetCbufU32) {
+        inst.ReplaceUsesWith(IR::Value{env.ReadCbufValue(bank_value, offset_value)});
+    } else {
+        inst.ReplaceUsesWith(
+            IR::Value{Common::BitCast<f32>(env.ReadCbufValue(bank_value, offset_value))});
+    }
+}
+
 void ConstantPropagation(Environment& env, IR::Block& block, IR::Inst& inst) {
     switch (inst.GetOpcode()) {
     case IR::Opcode::GetRegister:
@@ -825,13 +849,17 @@ void ConstantPropagation(Environment& env, IR::Block& block, IR::Inst& inst) {
     case IR::Opcode::GetCbufF32:
     case IR::Opcode::GetCbufU32:
         if (env.HasHLEMacroState()) {
-            return FoldConstBuffer(env, block, inst);
+            FoldConstBuffer(env, block, inst);
+        }
+        if (env.IsPropietaryDriver()) {
+            FoldDriverConstBuffer(env, block, inst, 1);
         }
         break;
     default:
         break;
     }
 }
+
 } // Anonymous namespace
 
 void ConstantPropagationPass(Environment& env, IR::Program& program) {
diff --git a/src/video_core/renderer_opengl/gl_shader_cache.cpp b/src/video_core/renderer_opengl/gl_shader_cache.cpp
index bf991afee..03b6314ff 100644
--- a/src/video_core/renderer_opengl/gl_shader_cache.cpp
+++ b/src/video_core/renderer_opengl/gl_shader_cache.cpp
@@ -51,7 +51,7 @@ using VideoCommon::LoadPipelines;
 using VideoCommon::SerializePipeline;
 using Context = ShaderContext::Context;
 
-constexpr u32 CACHE_VERSION = 8;
+constexpr u32 CACHE_VERSION = 9;
 
 template <typename Container>
 auto MakeSpan(Container& container) {
diff --git a/src/video_core/renderer_vulkan/vk_pipeline_cache.cpp b/src/video_core/renderer_vulkan/vk_pipeline_cache.cpp
index 6cd162422..3046b72ab 100644
--- a/src/video_core/renderer_vulkan/vk_pipeline_cache.cpp
+++ b/src/video_core/renderer_vulkan/vk_pipeline_cache.cpp
@@ -54,7 +54,7 @@ using VideoCommon::FileEnvironment;
 using VideoCommon::GenericEnvironment;
 using VideoCommon::GraphicsEnvironment;
 
-constexpr u32 CACHE_VERSION = 9;
+constexpr u32 CACHE_VERSION = 10;
 
 template <typename Container>
 auto MakeSpan(Container& container) {
diff --git a/src/video_core/shader_environment.cpp b/src/video_core/shader_environment.cpp
index 99d85bfb3..c34728245 100644
--- a/src/video_core/shader_environment.cpp
+++ b/src/video_core/shader_environment.cpp
@@ -325,6 +325,7 @@ GraphicsEnvironment::GraphicsEnvironment(Tegra::Engines::Maxwell3D& maxwell3d_,
     ASSERT(local_size <= std::numeric_limits<u32>::max());
     local_memory_size = static_cast<u32>(local_size) + sph.common3.shader_local_memory_crs_size;
     texture_bound = maxwell3d->regs.bindless_texture_const_buffer_slot;
+    is_propietary_driver = texture_bound == 2;
     has_hle_engine_state =
         maxwell3d->engine_state == Tegra::Engines::Maxwell3D::EngineHint::OnHLEMacro;
 }
@@ -399,6 +400,7 @@ ComputeEnvironment::ComputeEnvironment(Tegra::Engines::KeplerCompute& kepler_com
     stage = Shader::Stage::Compute;
     local_memory_size = qmd.local_pos_alloc + qmd.local_crs_alloc;
     texture_bound = kepler_compute->regs.tex_cb_index;
+    is_propietary_driver = texture_bound == 2;
     shared_memory_size = qmd.shared_alloc;
     workgroup_size = {qmd.block_dim_x, qmd.block_dim_y, qmd.block_dim_z};
 }
@@ -498,6 +500,7 @@ void FileEnvironment::Deserialize(std::ifstream& file) {
             file.read(reinterpret_cast<char*>(&gp_passthrough_mask), sizeof(gp_passthrough_mask));
         }
     }
+    is_propietary_driver = texture_bound == 2;
 }
 
 void FileEnvironment::Dump(u64 hash) {