diff options
| -rw-r--r-- | src/shader_recompiler/environment.h | 5 | ||||
| -rw-r--r-- | src/shader_recompiler/ir_opt/constant_propagation_pass.cpp | 30 | ||||
| -rw-r--r-- | src/video_core/renderer_opengl/gl_shader_cache.cpp | 2 | ||||
| -rw-r--r-- | src/video_core/renderer_vulkan/vk_pipeline_cache.cpp | 2 | ||||
| -rw-r--r-- | src/video_core/shader_environment.cpp | 3 |
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) { |