GPU processor refactoring (#1787)

* coroutine code prettification

* asc queues submission refactoring

* better asc ring context handling

* final touches and review notes

* even more simplification for context saving

src/video_core/renderer_vulkan/vk_pipeline_cache.cpp

@@ -173,9 +173,9 @@ Shader::RuntimeInfo PipelineCache::BuildRuntimeInfo(Stage stage, LogicalStage l_
         break;
     }
     case Stage::Compute: {
-        const auto& cs_pgm = regs.cs_program;
+        const auto& cs_pgm = liverpool->GetCsRegs();
         info.num_user_data = cs_pgm.settings.num_user_regs;
-        info.num_allocated_vgprs = regs.cs_program.settings.num_vgprs * 4;
+        info.num_allocated_vgprs = cs_pgm.settings.num_vgprs * 4;
         info.cs_info.workgroup_size = {cs_pgm.num_thread_x.full, cs_pgm.num_thread_y.full,
                                        cs_pgm.num_thread_z.full};
         info.cs_info.tgid_enable = {cs_pgm.IsTgidEnabled(0), cs_pgm.IsTgidEnabled(1),
@@ -476,8 +476,8 @@ bool PipelineCache::RefreshGraphicsKey() {
 
 bool PipelineCache::RefreshComputeKey() {
     Shader::Backend::Bindings binding{};
-    const auto* cs_pgm = &liverpool->regs.cs_program;
-    const auto cs_params = Liverpool::GetParams(*cs_pgm);
+    const auto& cs_pgm = liverpool->GetCsRegs();
+    const auto cs_params = Liverpool::GetParams(cs_pgm);
     std::tie(infos[0], modules[0], fetch_shader, compute_key.value) =
         GetProgram(Shader::Stage::Compute, LogicalStage::Compute, cs_params, binding);
     return true;
@@ -529,6 +529,7 @@ PipelineCache::Result PipelineCache::GetProgram(Stage stage, LogicalStage l_stag
         return std::make_tuple(&program->info, module, spec.fetch_shader_data,
                                HashCombine(params.hash, 0));
     }
+    it_pgm.value()->info.user_data = params.user_data;
 
     auto& program = it_pgm.value();
     auto& info = program->info;

src/video_core/renderer_vulkan/vk_rasterizer.cpp

@@ -317,14 +317,14 @@ void Rasterizer::DrawIndirect(bool is_indexed, VAddr arg_address, u32 offset, u3
 void Rasterizer::DispatchDirect() {
     RENDERER_TRACE;
 
-    const auto& cs_program = liverpool->regs.cs_program;
+    const auto& cs_program = liverpool->GetCsRegs();
     const ComputePipeline* pipeline = pipeline_cache.GetComputePipeline();
     if (!pipeline) {
         return;
     }
 
     const auto& cs = pipeline->GetStage(Shader::LogicalStage::Compute);
-    if (ExecuteShaderHLE(cs, liverpool->regs, *this)) {
+    if (ExecuteShaderHLE(cs, liverpool->regs, cs_program, *this)) {
         return;
     }
 
@@ -344,7 +344,7 @@ void Rasterizer::DispatchDirect() {
 void Rasterizer::DispatchIndirect(VAddr address, u32 offset, u32 size) {
     RENDERER_TRACE;
 
-    const auto& cs_program = liverpool->regs.cs_program;
+    const auto& cs_program = liverpool->GetCsRegs();
     const ComputePipeline* pipeline = pipeline_cache.GetComputePipeline();
     if (!pipeline) {
         return;

src/video_core/renderer_vulkan/vk_shader_hle.cpp

@@ -2,17 +2,19 @@
 // SPDX-License-Identifier: GPL-2.0-or-later
 
 #include "shader_recompiler/info.h"
+#include "video_core/renderer_vulkan/vk_rasterizer.h"
 #include "video_core/renderer_vulkan/vk_scheduler.h"
 #include "video_core/renderer_vulkan/vk_shader_hle.h"
 
-#include "vk_rasterizer.h"
+extern std::unique_ptr<AmdGpu::Liverpool> liverpool;
 
 namespace Vulkan {
 
 static constexpr u64 COPY_SHADER_HASH = 0xfefebf9f;
 
-bool ExecuteCopyShaderHLE(const Shader::Info& info, const AmdGpu::Liverpool::Regs& regs,
-                          Rasterizer& rasterizer) {
+static bool ExecuteCopyShaderHLE(const Shader::Info& info,
+                                 const AmdGpu::Liverpool::ComputeProgram& cs_program,
+                                 Rasterizer& rasterizer) {
     auto& scheduler = rasterizer.GetScheduler();
     auto& buffer_cache = rasterizer.GetBufferCache();
 
@@ -34,9 +36,9 @@ bool ExecuteCopyShaderHLE(const Shader::Info& info, const AmdGpu::Liverpool::Reg
 
     static std::vector<vk::BufferCopy> copies;
     copies.clear();
-    copies.reserve(regs.cs_program.dim_x);
+    copies.reserve(cs_program.dim_x);
 
-    for (u32 i = 0; i < regs.cs_program.dim_x; i++) {
+    for (u32 i = 0; i < cs_program.dim_x; i++) {
         const auto& [dst_idx, src_idx, end] = ctl_buf[i];
         const u32 local_dst_offset = dst_idx * buf_stride;
         const u32 local_src_offset = src_idx * buf_stride;
@@ -122,10 +124,10 @@ bool ExecuteCopyShaderHLE(const Shader::Info& info, const AmdGpu::Liverpool::Reg
 }
 
 bool ExecuteShaderHLE(const Shader::Info& info, const AmdGpu::Liverpool::Regs& regs,
-                      Rasterizer& rasterizer) {
+                      const AmdGpu::Liverpool::ComputeProgram& cs_program, Rasterizer& rasterizer) {
     switch (info.pgm_hash) {
     case COPY_SHADER_HASH:
-        return ExecuteCopyShaderHLE(info, regs, rasterizer);
+        return ExecuteCopyShaderHLE(info, cs_program, rasterizer);
     default:
         return false;
     }

src/video_core/renderer_vulkan/vk_shader_hle.h

@@ -15,6 +15,6 @@ class Rasterizer;
 
 /// Attempts to execute a shader using HLE if possible.
 bool ExecuteShaderHLE(const Shader::Info& info, const AmdGpu::Liverpool::Regs& regs,
-                      Rasterizer& rasterizer);
+                      const AmdGpu::Liverpool::ComputeProgram& cs_program, Rasterizer& rasterizer);
 
 } // namespace Vulkan