I'm trying to render grass using compute shader evaluations (wind, recovery and gravity). So I implemented indirect rendering and decided to store a separate compute command buffer. This is how I record compute buffer:
void record_compute_command_buffer() {
vk::CommandBufferBeginInfo begin_info{};
begin_info.flags = vk::CommandBufferUsageFlagBits::eSimultaneousUse;
GPU_.compute_command_buffer_.begin(begin_info);
GPU_.compute_command_buffer_.bindPipeline(vk::PipelineBindPoint::eCompute, GPU_.compute_pipeline_);
blade_compute_push_data push{
camera_.get_view(),
glm::perspective(glm::radians(90.0f), GPU_.aspect_ratio(), 0.1f, 10.0f),
time_.delta_time,
time_.total_time
};
push.projection_matrix[1][1] *= -1;
GPU_.compute_command_buffer_.pushConstants(GPU_.compute_pipeline_layout_, vk::ShaderStageFlagBits::eCompute, 0, sizeof(push), &push);
GPU_.compute_command_buffer_.bindDescriptorSets(
vk::PipelineBindPoint::eCompute,
GPU_.compute_pipeline_layout_,
0,
1,
&GPU_.compute_descriptor_sets_[0],
0,
nullptr
);
const int workgroup_size = 32;
const int groupcount = ((blades.size()) / workgroup_size) + 1;
GPU_.compute_command_buffer_.dispatch(groupcount, 1, 1);
GPU_.compute_command_buffer_.end();
}
If I create 31 blades it all goes fine. But if the blades' size is more than 31 it all results in this (video): https://drive.google.com/drive/folders/1pYByX6qWQUiRbvDldOgTJH6scrwM5J1S
And I get no validation layer errors.
My compute shader:
#version 450
layout(local_size_x = 32) in;
layout(push_constant) uniform push_data {
mat4 view;
mat4 proj;
float delta_time;
float total_time;
} push;
struct blade_t {
vec4 v0;
vec4 v1;
vec4 v2;
vec4 up;
};
layout(set = 0, binding = 0) buffer input_blades {
blade_t all_blades[];
};
layout(set = 0, binding = 1) buffer culled_blades {
blade_t result[];
};
layout(set = 0, binding = 2) buffer indirect_draw_params {
uint vertex_count; // keeps updating
uint instance_count; // 1
uint first_vertex; // 0
uint first_instance; // 0
} indirect_params;
void main() {
uint id = gl_GlobalInvocationID.x;
if (id > indirect_params.vertex_count) return;
if (id == 0) indirect_params.vertex_count = 0;
barrier();
blade_t cur_blade = all_blades[id];
result[atomicAdd(indirect_params.vertex_count, 1)] = cur_blade;
}
if (id >= all_blades.length()) returnstatement solved the problem. Modifiedmainfunction in compute shader: