Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Fix indirect render using the wrong index #404

Merged
merged 1 commit into from
Dec 7, 2024
Merged
Show file tree
Hide file tree
Changes from all commits
Commits
File filter

Filter by extension

Filter by extension

Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
12 changes: 12 additions & 0 deletions src/render/mod.rs
Original file line number Diff line number Diff line change
Expand Up @@ -378,6 +378,9 @@ pub struct GpuParticleGroup {
/// Index of the [`GpuDispatchIndirect`] struct inside the global
/// [`EffectsMeta::dispatch_indirect_buffer`].
pub indirect_dispatch_index: u32,
/// Index of the [`GpuRenderGroupIndirect`] struct inside the global
/// [`EffectsMeta::render_group_dispatch_buffer`].
pub indirect_render_index: u32,
}

/// Compute pipeline to run the `vfx_indirect` dispatch workgroup calculation
Expand Down Expand Up @@ -2434,6 +2437,14 @@ pub(crate) fn prepare_effects(
let mut first_particle_group_buffer_index = None;
let mut local_group_count = 0;
for (group_index, range) in input.effect_slices.slices.windows(2).enumerate() {
let indirect_render_index = match &dispatch_buffer_indices.render_group_dispatch_indices
{
RenderGroupDispatchIndices::Allocated {
first_render_group_dispatch_buffer_index,
..
} => first_render_group_dispatch_buffer_index.0 + group_index as u32,
_ => u32::MAX, // should never happen, as lazily allocated above (unless something went wrong)
};
let particle_group_buffer_index =
effects_meta.particle_group_buffer.push(GpuParticleGroup {
global_group_index: total_group_count,
Expand All @@ -2448,6 +2459,7 @@ pub(crate) fn prepare_effects(
.first_update_group_dispatch_buffer_index
.0
+ group_index as u32,
indirect_render_index,
});
if group_index == 0 {
first_particle_group_buffer_index = Some(particle_group_buffer_index as u32);
Expand Down
3 changes: 3 additions & 0 deletions src/render/vfx_common.wgsl
Original file line number Diff line number Diff line change
Expand Up @@ -71,6 +71,9 @@ struct ParticleGroup {
/// Index of the [`GpuDispatchIndirect`] struct inside the global
/// [`EffectsMeta::dispatch_indirect_buffer`].
indirect_dispatch_index: u32,
/// Index of the [`GpuRenderGroupIndirect`] struct inside the global
/// [`EffectsMeta::render_group_dispatch_buffer`].
indirect_render_index: u32,
{{PARTICLE_GROUP_PADDING}}
}

Expand Down
11 changes: 4 additions & 7 deletions src/render/vfx_indirect.wgsl
Original file line number Diff line number Diff line change
Expand Up @@ -43,15 +43,10 @@ fn main(@builtin(global_invocation_id) global_invocation_id: vec3<u32>) {
let group_index_in_effect = group_buffer[index].index_in_effect;
let is_first_group = group_index_in_effect == 0u;

// Calculate the base offset (in number of u32 items) into the render indirect and
// dispatch indirect arrays.
let rgi_base = RENDER_GROUP_INDIRECT_STRIDE * group_index;

let indirect_dispatch_index = group_buffer[index].indirect_dispatch_index;
let di_base = DISPATCH_INDIRECT_STRIDE * indirect_dispatch_index;

// Clear the rendering instance count, which will be upgraded by the update pass
// with the particles actually alive at the end of their update (after aged).
let indirect_render_index = group_buffer[index].indirect_render_index;
let rgi_base = RENDER_GROUP_INDIRECT_STRIDE * indirect_render_index;
render_group_indirect_buffer[rgi_base + RGI_OFFSET_INSTANCE_COUNT] = 0u;

let alive_count = render_group_indirect_buffer[rgi_base + RGI_OFFSET_ALIVE_COUNT];
Expand All @@ -60,6 +55,8 @@ fn main(@builtin(global_invocation_id) global_invocation_id: vec3<u32>) {
// Calculate the number of thread groups to dispatch for the update
// pass, which is the number of alive particles rounded up to 64
// (workgroup_size).
let indirect_dispatch_index = group_buffer[index].indirect_dispatch_index;
let di_base = DISPATCH_INDIRECT_STRIDE * indirect_dispatch_index;
dispatch_indirect_buffer[di_base + DI_OFFSET_X] = (alive_count + 63u) >> 6u;

// Update max_update from current value of alive_count, so that the
Expand Down
Loading