// Copyright 2021 yuzu Emulator Project // Licensed under GPLv2 or any later version // Refer to the license.txt file included. #include #include #include #include "shader_recompiler/frontend/ir/basic_block.h" #include "shader_recompiler/frontend/ir/post_order.h" #include "shader_recompiler/frontend/maxwell/program.h" #include "shader_recompiler/frontend/maxwell/structured_control_flow.h" #include "shader_recompiler/frontend/maxwell/translate/translate.h" #include "shader_recompiler/ir_opt/passes.h" namespace Shader::Maxwell { namespace { void RemoveUnreachableBlocks(IR::Program& program) { // Some blocks might be unreachable if a function call exists unconditionally // If this happens the number of blocks and post order blocks will mismatch if (program.blocks.size() == program.post_order_blocks.size()) { return; } const auto begin{program.blocks.begin() + 1}; const auto end{program.blocks.end()}; const auto pred{[](IR::Block* block) { return block->ImmediatePredecessors().empty(); }}; program.blocks.erase(std::remove_if(begin, end, pred), end); } void CollectInterpolationInfo(Environment& env, IR::Program& program) { if (program.stage != Stage::Fragment) { return; } const ProgramHeader& sph{env.SPH()}; for (size_t index = 0; index < program.info.input_generics.size(); ++index) { std::optional imap; for (const PixelImap value : sph.ps.GenericInputMap(static_cast(index))) { if (value == PixelImap::Unused) { continue; } if (imap && imap != value) { throw NotImplementedException("Per component interpolation"); } imap = value; } if (!imap) { continue; } program.info.input_generics[index].interpolation = [&] { switch (*imap) { case PixelImap::Unused: case PixelImap::Perspective: return Interpolation::Smooth; case PixelImap::Constant: return Interpolation::Flat; case PixelImap::ScreenLinear: return Interpolation::NoPerspective; } throw NotImplementedException("Unknown interpolation {}", *imap); }(); } } void AddNVNStorageBuffers(IR::Program& program) { if (!program.info.uses_global_memory) { return; } const u32 driver_cbuf{0}; const u32 descriptor_size{0x10}; const u32 num_buffers{16}; const u32 base{[&] { switch (program.stage) { case Stage::VertexA: case Stage::VertexB: return 0x110u; case Stage::TessellationControl: return 0x210u; case Stage::TessellationEval: return 0x310u; case Stage::Geometry: return 0x410u; case Stage::Fragment: return 0x510u; case Stage::Compute: return 0x310u; } throw InvalidArgument("Invalid stage {}", program.stage); }()}; auto& descs{program.info.storage_buffers_descriptors}; for (u32 index = 0; index < num_buffers; ++index) { const u32 offset{base + index * descriptor_size}; const auto it{std::ranges::find(descs, offset, &StorageBufferDescriptor::cbuf_offset)}; if (it != descs.end()) { continue; } // Assume these are written for now descs.push_back({ .cbuf_index = driver_cbuf, .cbuf_offset = offset, .count = 1, .is_written = true, }); } } } // Anonymous namespace IR::Program TranslateProgram(ObjectPool& inst_pool, ObjectPool& block_pool, Environment& env, Flow::CFG& cfg) { IR::Program program; program.blocks = VisitAST(inst_pool, block_pool, env, cfg); program.post_order_blocks = PostOrder(program.blocks); program.stage = env.ShaderStage(); program.local_memory_size = env.LocalMemorySize(); switch (program.stage) { case Stage::TessellationControl: { const ProgramHeader& sph{env.SPH()}; program.invocations = sph.common2.threads_per_input_primitive; break; } case Stage::Geometry: { const ProgramHeader& sph{env.SPH()}; program.output_topology = sph.common3.output_topology; program.output_vertices = sph.common4.max_output_vertices; program.invocations = sph.common2.threads_per_input_primitive; break; } case Stage::Compute: program.workgroup_size = env.WorkgroupSize(); program.shared_memory_size = env.SharedMemorySize(); break; default: break; } RemoveUnreachableBlocks(program); // Replace instructions before the SSA rewrite Optimization::LowerFp16ToFp32(program); Optimization::SsaRewritePass(program); Optimization::GlobalMemoryToStorageBufferPass(program); Optimization::TexturePass(env, program); Optimization::ConstantPropagationPass(program); Optimization::DeadCodeEliminationPass(program); Optimization::VerificationPass(program); Optimization::CollectShaderInfoPass(env, program); CollectInterpolationInfo(env, program); AddNVNStorageBuffers(program); return program; } } // namespace Shader::Maxwell