2fd4cc4937
* Use texture atomics rather than buffer atomics for the visbuffer (haven't tested perf on a raster-heavy scene yet) * Unfortunately to clear the visbuffer we now need a compute pass to clear it. Using wgpu's clear_texture function internally uses a buffer -> image copy that's insanely expensive. Ideally it should be using vkCmdClearColorImage, which I've opened an issue for https://github.com/gfx-rs/wgpu/issues/7090. For now we'll have to stick with a custom compute pass and all the extra code that brings. * Faster resolve depth pass by discarding 0 depth pixels instead of redundantly writing zero (2x faster for big depth textures like shadow views)
19 lines
660 B
WebGPU Shading Language
19 lines
660 B
WebGPU Shading Language
#ifdef MESHLET_VISIBILITY_BUFFER_RASTER_PASS_OUTPUT
|
|
@group(0) @binding(0) var meshlet_visibility_buffer: texture_storage_2d<r64uint, write>;
|
|
#else
|
|
@group(0) @binding(0) var meshlet_visibility_buffer: texture_storage_2d<r32uint, write>;
|
|
#endif
|
|
var<push_constant> view_size: vec2<u32>;
|
|
|
|
@compute
|
|
@workgroup_size(16, 16, 1)
|
|
fn clear_visibility_buffer(@builtin(global_invocation_id) global_id: vec3<u32>) {
|
|
if any(global_id.xy >= view_size) { return; }
|
|
|
|
#ifdef MESHLET_VISIBILITY_BUFFER_RASTER_PASS_OUTPUT
|
|
textureStore(meshlet_visibility_buffer, global_id.xy, vec4(0lu));
|
|
#else
|
|
textureStore(meshlet_visibility_buffer, global_id.xy, vec4(0u));
|
|
#endif
|
|
}
|