From 64337f004d9249c4408fec75bd1bbcc0f2a1408d Mon Sep 17 00:00:00 2001
From: ameerj <52414509+ameerj@users.noreply.github.com>
Date: Thu, 20 May 2021 23:38:38 -0400
Subject: glsl: Fix "reg" allocing
based on glasm with some tweaks
---
.../backend/glsl/emit_context.cpp | 28 ++++++++++++++++++----
1 file changed, 24 insertions(+), 4 deletions(-)
(limited to 'src/shader_recompiler/backend/glsl/emit_context.cpp')
diff --git a/src/shader_recompiler/backend/glsl/emit_context.cpp b/src/shader_recompiler/backend/glsl/emit_context.cpp
index e2a9885f0d..b0e46cc07e 100644
--- a/src/shader_recompiler/backend/glsl/emit_context.cpp
+++ b/src/shader_recompiler/backend/glsl/emit_context.cpp
@@ -11,19 +11,39 @@ namespace Shader::Backend::GLSL {
EmitContext::EmitContext(IR::Program& program, [[maybe_unused]] Bindings& bindings,
const Profile& profile_)
: info{program.info}, profile{profile_} {
- std::string header = "#version 450 core\n";
- header += "layout(local_size_x=1, local_size_y=1, local_size_z=1) in;";
+ std::string header = "#version 450\n";
+ if (program.stage == Stage::Compute) {
+ header += fmt::format("layout(local_size_x={},local_size_y={},local_size_z={}) in;\n",
+ program.workgroup_size[0], program.workgroup_size[1],
+ program.workgroup_size[2]);
+ }
code += header;
DefineConstantBuffers();
- code += "void main(){";
+ DefineStorageBuffers();
+ code += "void main(){\n";
}
void EmitContext::DefineConstantBuffers() {
if (info.constant_buffer_descriptors.empty()) {
return;
}
+ u32 binding{};
for (const auto& desc : info.constant_buffer_descriptors) {
- Add("uniform uint c{}[{}];", desc.index, desc.count);
+ Add("layout(std140,binding={}) uniform cbuf_{}{{uint cbuf{}[];}};", binding, binding,
+ desc.index, desc.count);
+ ++binding;
+ }
+}
+
+void EmitContext::DefineStorageBuffers() {
+ if (info.storage_buffers_descriptors.empty()) {
+ return;
+ }
+ u32 binding{};
+ for (const auto& desc : info.storage_buffers_descriptors) {
+ Add("layout(std430,binding={}) buffer buff_{}{{uint buff{}[];}};", binding, binding,
+ desc.cbuf_index, desc.count);
+ ++binding;
}
}
--
cgit v1.2.3-70-g09d2