From 376aa94819b7da976adb120136d83980a757d044 Mon Sep 17 00:00:00 2001 From: ReinUsesLisp Date: Wed, 16 Jun 2021 01:49:19 -0300 Subject: shader: Rename maxwell/program.h to translate_program.h --- src/shader_recompiler/CMakeLists.txt | 4 +- src/shader_recompiler/frontend/maxwell/program.cpp | 203 --------------------- src/shader_recompiler/frontend/maxwell/program.h | 27 --- .../frontend/maxwell/translate_program.cpp | 203 +++++++++++++++++++++ .../frontend/maxwell/translate_program.h | 22 +++ src/video_core/renderer_opengl/gl_shader_cache.cpp | 2 +- .../renderer_vulkan/vk_pipeline_cache.cpp | 2 +- 7 files changed, 229 insertions(+), 234 deletions(-) delete mode 100644 src/shader_recompiler/frontend/maxwell/program.cpp delete mode 100644 src/shader_recompiler/frontend/maxwell/program.h create mode 100644 src/shader_recompiler/frontend/maxwell/translate_program.cpp create mode 100644 src/shader_recompiler/frontend/maxwell/translate_program.h diff --git a/src/shader_recompiler/CMakeLists.txt b/src/shader_recompiler/CMakeLists.txt index 06ee50fff..f801a9f72 100644 --- a/src/shader_recompiler/CMakeLists.txt +++ b/src/shader_recompiler/CMakeLists.txt @@ -114,8 +114,6 @@ add_library(shader_recompiler STATIC frontend/maxwell/maxwell.inc frontend/maxwell/opcodes.cpp frontend/maxwell/opcodes.h - frontend/maxwell/program.cpp - frontend/maxwell/program.h frontend/maxwell/structured_control_flow.cpp frontend/maxwell/structured_control_flow.h frontend/maxwell/translate/impl/atomic_operations_global_memory.cpp @@ -211,6 +209,8 @@ add_library(shader_recompiler STATIC frontend/maxwell/translate/impl/warp_shuffle.cpp frontend/maxwell/translate/translate.cpp frontend/maxwell/translate/translate.h + frontend/maxwell/translate_program.cpp + frontend/maxwell/translate_program.h ir_opt/collect_shader_info_pass.cpp ir_opt/constant_propagation_pass.cpp ir_opt/dead_code_elimination_pass.cpp diff --git a/src/shader_recompiler/frontend/maxwell/program.cpp b/src/shader_recompiler/frontend/maxwell/program.cpp deleted file mode 100644 index 8489f9a5f..000000000 --- a/src/shader_recompiler/frontend/maxwell/program.cpp +++ /dev/null @@ -1,203 +0,0 @@ -// Copyright 2021 yuzu Emulator Project -// Licensed under GPLv2 or any later version -// Refer to the license.txt file included. - -#include -#include -#include -#include - -#include "shader_recompiler/exception.h" -#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 { -IR::BlockList GenerateBlocks(const IR::AbstractSyntaxList& syntax_list) { - auto syntax_blocks{syntax_list | std::views::filter([](const auto& node) { - return node.type == IR::AbstractSyntaxNode::Type::Block; - })}; - IR::BlockList blocks(std::ranges::distance(syntax_blocks)); - std::ranges::transform(syntax_blocks, blocks.begin(), - [](const IR::AbstractSyntaxNode& node) { return node.data.block; }); - return blocks; -} - -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->ImmPredecessors().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) { - if (!program.info.nvn_buffer_used[index]) { - continue; - } - const u32 offset{base + index * descriptor_size}; - const auto it{std::ranges::find(descs, offset, &StorageBufferDescriptor::cbuf_offset)}; - if (it != descs.end()) { - it->is_written |= program.info.stores_global_memory; - continue; - } - descs.push_back({ - .cbuf_index = driver_cbuf, - .cbuf_offset = offset, - .count = 1, - .is_written = program.info.stores_global_memory, - }); - } -} -} // Anonymous namespace - -IR::Program TranslateProgram(ObjectPool& inst_pool, ObjectPool& block_pool, - Environment& env, Flow::CFG& cfg) { - IR::Program program; - program.syntax_list = BuildASL(inst_pool, block_pool, env, cfg); - program.blocks = GenerateBlocks(program.syntax_list); - program.post_order_blocks = PostOrder(program.syntax_list.front()); - 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; -} - -IR::Program MergeDualVertexPrograms(IR::Program& vertex_a, IR::Program& vertex_b, - Environment& env_vertex_b) { - IR::Program result{}; - Optimization::VertexATransformPass(vertex_a); - Optimization::VertexBTransformPass(vertex_b); - for (const auto& term : vertex_a.syntax_list) { - if (term.type == IR::AbstractSyntaxNode::Type::Return) { - continue; - } - result.syntax_list.push_back(term); - } - for (const auto& term : vertex_b.syntax_list) { - result.syntax_list.push_back(term); - } - result.blocks = GenerateBlocks(result.syntax_list); - result.post_order_blocks = vertex_b.post_order_blocks; - for (const auto& block : vertex_a.post_order_blocks) { - result.post_order_blocks.push_back(block); - } - result.stage = Stage::VertexB; - result.info = vertex_a.info; - result.local_memory_size = std::max(vertex_a.local_memory_size, vertex_b.local_memory_size); - for (size_t index = 0; index < 32; ++index) { - result.info.input_generics[index].used |= vertex_b.info.input_generics[index].used; - result.info.stores_generics[index] |= vertex_b.info.stores_generics[index]; - } - Optimization::JoinTextureInfo(result.info, vertex_b.info); - Optimization::JoinStorageInfo(result.info, vertex_b.info); - Optimization::DeadCodeEliminationPass(result); - Optimization::VerificationPass(result); - Optimization::CollectShaderInfoPass(env_vertex_b, result); - return result; -} - -} // namespace Shader::Maxwell diff --git a/src/shader_recompiler/frontend/maxwell/program.h b/src/shader_recompiler/frontend/maxwell/program.h deleted file mode 100644 index f7f5930e4..000000000 --- a/src/shader_recompiler/frontend/maxwell/program.h +++ /dev/null @@ -1,27 +0,0 @@ -// Copyright 2021 yuzu Emulator Project -// Licensed under GPLv2 or any later version -// Refer to the license.txt file included. - -#pragma once - -#include -#include -#include - -#include - -#include "shader_recompiler/environment.h" -#include "shader_recompiler/frontend/ir/program.h" -#include "shader_recompiler/frontend/maxwell/control_flow.h" -#include "shader_recompiler/object_pool.h" - -namespace Shader::Maxwell { - -[[nodiscard]] IR::Program TranslateProgram(ObjectPool& inst_pool, - ObjectPool& block_pool, Environment& env, - Flow::CFG& cfg); - -[[nodiscard]] IR::Program MergeDualVertexPrograms(IR::Program& vertex_a, IR::Program& vertex_b, - Environment& env_vertex_b); - -} // namespace Shader::Maxwell diff --git a/src/shader_recompiler/frontend/maxwell/translate_program.cpp b/src/shader_recompiler/frontend/maxwell/translate_program.cpp new file mode 100644 index 000000000..e52170e3e --- /dev/null +++ b/src/shader_recompiler/frontend/maxwell/translate_program.cpp @@ -0,0 +1,203 @@ +// Copyright 2021 yuzu Emulator Project +// Licensed under GPLv2 or any later version +// Refer to the license.txt file included. + +#include +#include +#include +#include + +#include "shader_recompiler/exception.h" +#include "shader_recompiler/frontend/ir/basic_block.h" +#include "shader_recompiler/frontend/ir/post_order.h" +#include "shader_recompiler/frontend/maxwell/structured_control_flow.h" +#include "shader_recompiler/frontend/maxwell/translate/translate.h" +#include "shader_recompiler/frontend/maxwell/translate_program.h" +#include "shader_recompiler/ir_opt/passes.h" + +namespace Shader::Maxwell { +namespace { +IR::BlockList GenerateBlocks(const IR::AbstractSyntaxList& syntax_list) { + auto syntax_blocks{syntax_list | std::views::filter([](const auto& node) { + return node.type == IR::AbstractSyntaxNode::Type::Block; + })}; + IR::BlockList blocks(std::ranges::distance(syntax_blocks)); + std::ranges::transform(syntax_blocks, blocks.begin(), + [](const IR::AbstractSyntaxNode& node) { return node.data.block; }); + return blocks; +} + +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->ImmPredecessors().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) { + if (!program.info.nvn_buffer_used[index]) { + continue; + } + const u32 offset{base + index * descriptor_size}; + const auto it{std::ranges::find(descs, offset, &StorageBufferDescriptor::cbuf_offset)}; + if (it != descs.end()) { + it->is_written |= program.info.stores_global_memory; + continue; + } + descs.push_back({ + .cbuf_index = driver_cbuf, + .cbuf_offset = offset, + .count = 1, + .is_written = program.info.stores_global_memory, + }); + } +} +} // Anonymous namespace + +IR::Program TranslateProgram(ObjectPool& inst_pool, ObjectPool& block_pool, + Environment& env, Flow::CFG& cfg) { + IR::Program program; + program.syntax_list = BuildASL(inst_pool, block_pool, env, cfg); + program.blocks = GenerateBlocks(program.syntax_list); + program.post_order_blocks = PostOrder(program.syntax_list.front()); + 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; +} + +IR::Program MergeDualVertexPrograms(IR::Program& vertex_a, IR::Program& vertex_b, + Environment& env_vertex_b) { + IR::Program result{}; + Optimization::VertexATransformPass(vertex_a); + Optimization::VertexBTransformPass(vertex_b); + for (const auto& term : vertex_a.syntax_list) { + if (term.type == IR::AbstractSyntaxNode::Type::Return) { + continue; + } + result.syntax_list.push_back(term); + } + for (const auto& term : vertex_b.syntax_list) { + result.syntax_list.push_back(term); + } + result.blocks = GenerateBlocks(result.syntax_list); + result.post_order_blocks = vertex_b.post_order_blocks; + for (const auto& block : vertex_a.post_order_blocks) { + result.post_order_blocks.push_back(block); + } + result.stage = Stage::VertexB; + result.info = vertex_a.info; + result.local_memory_size = std::max(vertex_a.local_memory_size, vertex_b.local_memory_size); + for (size_t index = 0; index < 32; ++index) { + result.info.input_generics[index].used |= vertex_b.info.input_generics[index].used; + result.info.stores_generics[index] |= vertex_b.info.stores_generics[index]; + } + Optimization::JoinTextureInfo(result.info, vertex_b.info); + Optimization::JoinStorageInfo(result.info, vertex_b.info); + Optimization::DeadCodeEliminationPass(result); + Optimization::VerificationPass(result); + Optimization::CollectShaderInfoPass(env_vertex_b, result); + return result; +} + +} // namespace Shader::Maxwell diff --git a/src/shader_recompiler/frontend/maxwell/translate_program.h b/src/shader_recompiler/frontend/maxwell/translate_program.h new file mode 100644 index 000000000..1e5536443 --- /dev/null +++ b/src/shader_recompiler/frontend/maxwell/translate_program.h @@ -0,0 +1,22 @@ +// Copyright 2021 yuzu Emulator Project +// Licensed under GPLv2 or any later version +// Refer to the license.txt file included. + +#pragma once + +#include "shader_recompiler/environment.h" +#include "shader_recompiler/frontend/ir/basic_block.h" +#include "shader_recompiler/frontend/ir/program.h" +#include "shader_recompiler/frontend/maxwell/control_flow.h" +#include "shader_recompiler/object_pool.h" + +namespace Shader::Maxwell { + +[[nodiscard]] IR::Program TranslateProgram(ObjectPool& inst_pool, + ObjectPool& block_pool, Environment& env, + Flow::CFG& cfg); + +[[nodiscard]] IR::Program MergeDualVertexPrograms(IR::Program& vertex_a, IR::Program& vertex_b, + Environment& env_vertex_b); + +} // namespace Shader::Maxwell diff --git a/src/video_core/renderer_opengl/gl_shader_cache.cpp b/src/video_core/renderer_opengl/gl_shader_cache.cpp index 620666622..c05cd5d28 100644 --- a/src/video_core/renderer_opengl/gl_shader_cache.cpp +++ b/src/video_core/renderer_opengl/gl_shader_cache.cpp @@ -22,7 +22,7 @@ #include "shader_recompiler/backend/spirv/emit_spirv.h" #include "shader_recompiler/frontend/ir/program.h" #include "shader_recompiler/frontend/maxwell/control_flow.h" -#include "shader_recompiler/frontend/maxwell/program.h" +#include "shader_recompiler/frontend/maxwell/translate_program.h" #include "shader_recompiler/profile.h" #include "video_core/engines/kepler_compute.h" #include "video_core/engines/maxwell_3d.h" diff --git a/src/video_core/renderer_vulkan/vk_pipeline_cache.cpp b/src/video_core/renderer_vulkan/vk_pipeline_cache.cpp index b17f34cdd..0b6fe8e2e 100644 --- a/src/video_core/renderer_vulkan/vk_pipeline_cache.cpp +++ b/src/video_core/renderer_vulkan/vk_pipeline_cache.cpp @@ -20,7 +20,7 @@ #include "shader_recompiler/backend/spirv/emit_spirv.h" #include "shader_recompiler/environment.h" #include "shader_recompiler/frontend/maxwell/control_flow.h" -#include "shader_recompiler/frontend/maxwell/program.h" +#include "shader_recompiler/frontend/maxwell/translate_program.h" #include "shader_recompiler/program_header.h" #include "video_core/dirty_flags.h" #include "video_core/engines/kepler_compute.h" -- cgit v1.2.3