From af9c945f40a7df91365b446105a33b00f0930763 Mon Sep 17 00:00:00 2001 From: JMS55 <47158642+JMS55@users.noreply.github.com> Date: Sun, 18 Jun 2023 17:05:55 -0400 Subject: [PATCH] Screen Space Ambient Occlusion (SSAO) MVP (#7402) ![image](https://github.com/bevyengine/bevy/assets/47158642/dbb62645-f639-4f2b-b84b-26fd915c186d) # Objective - Add Screen space ambient occlusion (SSAO). SSAO approximates small-scale, local occlusion of _indirect_ diffuse light between objects. SSAO does not apply to direct lighting, such as point or directional lights. - This darkens creases, e.g. on staircases, and gives nice contact shadows where objects meet, giving entities a more "grounded" feel. - Closes https://github.com/bevyengine/bevy/issues/3632. ## Solution - Implement the GTAO algorithm. - https://www.activision.com/cdn/research/Practical_Real_Time_Strategies_for_Accurate_Indirect_Occlusion_NEW%20VERSION_COLOR.pdf - https://blog.selfshadow.com/publications/s2016-shading-course/activision/s2016_pbs_activision_occlusion.pdf - Source code heavily based on [Intel's XeGTAO](https://github.com/GameTechDev/XeGTAO/blob/0d177ce06bfa642f64d8af4de1197ad1bcb862d4/Source/Rendering/Shaders/XeGTAO.hlsli). - Add an SSAO bevy example. ## Algorithm Overview * Run a depth and normal prepass * Create downscaled mips of the depth texture (preprocess_depths pass) * GTAO pass - for each pixel, take several random samples from the depth+normal buffers, reconstruct world position, raytrace in screen space to estimate occlusion. Rather then doing completely random samples on a hemisphere, you choose random _slices_ of the hemisphere, and then can analytically compute the full occlusion of that slice. Also compute edges based on depth differences here. * Spatial denoise pass - bilateral blur, using edge detection to not blur over edges. This is the final SSAO result. * Main pass - if SSAO exists, sample the SSAO texture, and set occlusion to be the minimum of ssao/material occlusion. This then feeds into the rest of the PBR shader as normal. --- ## Future Improvements - Maybe remove the low quality preset for now (too noisy) - WebGPU fallback (see below) - Faster depth->world position (see reverted code) - Bent normals - Try interleaved gradient noise or spatiotemporal blue noise - Replace the spatial denoiser with a combined spatial+temporal denoiser - Render at half resolution and use a bilateral upsample - Better multibounce approximation (https://drive.google.com/file/d/1SyagcEVplIm2KkRD3WQYSO9O0Iyi1hfy/view) ## Far-Future Performance Improvements - F16 math (missing naga-wgsl support https://github.com/gfx-rs/naga/issues/1884) - Faster coordinate space conversion for normals - Faster depth mipchain creation (https://github.com/GPUOpen-Effects/FidelityFX-SPD) (wgpu/naga does not currently support subgroup ops) - Deinterleaved SSAO for better cache efficiency (https://developer.nvidia.com/sites/default/files/akamai/gameworks/samples/DeinterleavedTexturing.pdf) ## Other Interesting Papers - Visibility bitmask (https://link.springer.com/article/10.1007/s00371-022-02703-y, https://cdrinmatane.github.io/posts/cgspotlight-slides/) - Screen space diffuse lighting (https://github.com/Patapom/GodComplex/blob/master/Tests/TestHBIL/2018%20Mayaux%20-%20Horizon-Based%20Indirect%20Lighting%20(HBIL).pdf) ## Platform Support * SSAO currently does not work on DirectX12 due to issues with wgpu and naga: * https://github.com/gfx-rs/wgpu/pull/3798 * https://github.com/gfx-rs/naga/pull/2353 * SSAO currently does not work on WebGPU because r16float is not a valid storage texture format https://gpuweb.github.io/gpuweb/wgsl/#storage-texel-formats. We can fix this with a fallback to r32float. --- ## Changelog - Added ScreenSpaceAmbientOcclusionSettings, ScreenSpaceAmbientOcclusionQualityLevel, and ScreenSpaceAmbientOcclusionBundle --------- Co-authored-by: IceSentry Co-authored-by: IceSentry Co-authored-by: Daniel Chia Co-authored-by: Elabajaba Co-authored-by: Robert Swain Co-authored-by: robtfm <50659922+robtfm@users.noreply.github.com> Co-authored-by: Brandon Dyer Co-authored-by: Edgar Geier Co-authored-by: Nicola Papale Co-authored-by: Carter Anderson --- Cargo.toml | 10 + crates/bevy_core_pipeline/src/prepass/node.rs | 1 - crates/bevy_pbr/src/lib.rs | 4 + crates/bevy_pbr/src/material.rs | 10 +- crates/bevy_pbr/src/render/mesh.rs | 53 +- .../src/render/mesh_view_bindings.wgsl | 23 +- crates/bevy_pbr/src/render/pbr.wgsl | 16 +- crates/bevy_pbr/src/render/pbr_ambient.wgsl | 2 +- crates/bevy_pbr/src/render/pbr_functions.wgsl | 4 +- crates/bevy_pbr/src/render/utils.wgsl | 1 + crates/bevy_pbr/src/ssao/gtao.wgsl | 176 ++++ crates/bevy_pbr/src/ssao/gtao_utils.wgsl | 22 + crates/bevy_pbr/src/ssao/mod.rs | 989 ++++++++++++++++++ .../bevy_pbr/src/ssao/preprocess_depth.wgsl | 101 ++ crates/bevy_pbr/src/ssao/spatial_denoise.wgsl | 92 ++ crates/bevy_render/src/view/mod.rs | 4 +- examples/3d/ssao.rs | 199 ++++ examples/README.md | 1 + 18 files changed, 1678 insertions(+), 30 deletions(-) create mode 100644 crates/bevy_pbr/src/ssao/gtao.wgsl create mode 100644 crates/bevy_pbr/src/ssao/gtao_utils.wgsl create mode 100644 crates/bevy_pbr/src/ssao/mod.rs create mode 100644 crates/bevy_pbr/src/ssao/preprocess_depth.wgsl create mode 100644 crates/bevy_pbr/src/ssao/spatial_denoise.wgsl create mode 100644 examples/3d/ssao.rs diff --git a/Cargo.toml b/Cargo.toml index 355b0a52aed6c..522b31c73ca31 100644 --- a/Cargo.toml +++ b/Cargo.toml @@ -517,6 +517,16 @@ description = "Create a custom material to draw 3d lines" category = "3D Rendering" wasm = true +[[example]] +name = "ssao" +path = "examples/3d/ssao.rs" + +[package.metadata.example.ssao] +name = "Screen Space Ambient Occlusion" +description = "A scene showcasing screen space ambient occlusion" +category = "3D Rendering" +wasm = false + [[example]] name = "spotlight" path = "examples/3d/spotlight.rs" diff --git a/crates/bevy_core_pipeline/src/prepass/node.rs b/crates/bevy_core_pipeline/src/prepass/node.rs index 6d3502212d990..bdf997601372e 100644 --- a/crates/bevy_core_pipeline/src/prepass/node.rs +++ b/crates/bevy_core_pipeline/src/prepass/node.rs @@ -94,7 +94,6 @@ impl ViewNode for PrepassNode { stencil_ops: None, }), }); - if let Some(viewport) = camera.viewport.as_ref() { render_pass.set_camera_viewport(viewport); } diff --git a/crates/bevy_pbr/src/lib.rs b/crates/bevy_pbr/src/lib.rs index 9cbed8922c422..0f146ed121716 100644 --- a/crates/bevy_pbr/src/lib.rs +++ b/crates/bevy_pbr/src/lib.rs @@ -12,6 +12,7 @@ mod parallax; mod pbr_material; mod prepass; mod render; +mod ssao; pub use alpha::*; pub use bundle::*; @@ -23,6 +24,7 @@ pub use parallax::*; pub use pbr_material::*; pub use prepass::*; pub use render::*; +pub use ssao::*; pub mod prelude { #[doc(hidden)] @@ -38,6 +40,7 @@ pub mod prelude { material::{Material, MaterialPlugin}, parallax::ParallaxMappingMethod, pbr_material::StandardMaterial, + ssao::ScreenSpaceAmbientOcclusionPlugin, }; } @@ -184,6 +187,7 @@ impl Plugin for PbrPlugin { prepass_enabled: self.prepass_enabled, ..Default::default() }) + .add_plugin(ScreenSpaceAmbientOcclusionPlugin) .add_plugin(EnvironmentMapPlugin) .init_resource::() .init_resource::() diff --git a/crates/bevy_pbr/src/material.rs b/crates/bevy_pbr/src/material.rs index 0016634533eeb..559c880051f10 100644 --- a/crates/bevy_pbr/src/material.rs +++ b/crates/bevy_pbr/src/material.rs @@ -1,7 +1,7 @@ use crate::{ render, AlphaMode, DrawMesh, DrawPrepass, EnvironmentMapLight, MeshPipeline, MeshPipelineKey, - MeshUniform, PrepassPipelinePlugin, PrepassPlugin, RenderLightSystems, SetMeshBindGroup, - SetMeshViewBindGroup, Shadow, + MeshUniform, PrepassPipelinePlugin, PrepassPlugin, RenderLightSystems, + ScreenSpaceAmbientOcclusionSettings, SetMeshBindGroup, SetMeshViewBindGroup, Shadow, }; use bevy_app::{App, Plugin}; use bevy_asset::{AddAsset, AssetEvent, AssetServer, Assets, Handle}; @@ -387,6 +387,7 @@ pub fn queue_material_meshes( Option<&Tonemapping>, Option<&DebandDither>, Option<&EnvironmentMapLight>, + Option<&ScreenSpaceAmbientOcclusionSettings>, Option<&NormalPrepass>, Option<&TemporalAntiAliasSettings>, &mut RenderPhase, @@ -402,6 +403,7 @@ pub fn queue_material_meshes( tonemapping, dither, environment_map, + ssao, normal_prepass, taa_settings, mut opaque_phase, @@ -455,6 +457,10 @@ pub fn queue_material_meshes( } } + if ssao.is_some() { + view_key |= MeshPipelineKey::SCREEN_SPACE_AMBIENT_OCCLUSION; + } + let rangefinder = view.rangefinder3d(); for visible_entity in &visible_entities.entities { if let Ok((material_handle, mesh_handle, mesh_uniform)) = diff --git a/crates/bevy_pbr/src/render/mesh.rs b/crates/bevy_pbr/src/render/mesh.rs index d4a80c55acd32..89f95932ebc47 100644 --- a/crates/bevy_pbr/src/render/mesh.rs +++ b/crates/bevy_pbr/src/render/mesh.rs @@ -1,9 +1,9 @@ use crate::{ environment_map, prepass, EnvironmentMapLight, FogMeta, GlobalLightMeta, GpuFog, GpuLights, GpuPointLights, LightMeta, NotShadowCaster, NotShadowReceiver, PreviousGlobalTransform, - ShadowSamplers, ViewClusterBindings, ViewFogUniformOffset, ViewLightsUniformOffset, - ViewShadowBindings, CLUSTERED_FORWARD_STORAGE_BUFFER_COUNT, MAX_CASCADES_PER_LIGHT, - MAX_DIRECTIONAL_LIGHTS, + ScreenSpaceAmbientOcclusionTextures, ShadowSamplers, ViewClusterBindings, ViewFogUniformOffset, + ViewLightsUniformOffset, ViewShadowBindings, CLUSTERED_FORWARD_STORAGE_BUFFER_COUNT, + MAX_CASCADES_PER_LIGHT, MAX_DIRECTIONAL_LIGHTS, }; use bevy_app::Plugin; use bevy_asset::{load_internal_asset, Assets, Handle, HandleUntyped}; @@ -433,22 +433,33 @@ impl FromWorld for MeshPipeline { }, count: None, }, + // Screen space ambient occlusion texture + BindGroupLayoutEntry { + binding: 11, + visibility: ShaderStages::FRAGMENT, + ty: BindingType::Texture { + multisampled: false, + sample_type: TextureSampleType::Float { filterable: false }, + view_dimension: TextureViewDimension::D2, + }, + count: None, + }, ]; // EnvironmentMapLight let environment_map_entries = - environment_map::get_bind_group_layout_entries([11, 12, 13]); + environment_map::get_bind_group_layout_entries([12, 13, 14]); entries.extend_from_slice(&environment_map_entries); // Tonemapping - let tonemapping_lut_entries = get_lut_bind_group_layout_entries([14, 15]); + let tonemapping_lut_entries = get_lut_bind_group_layout_entries([15, 16]); entries.extend_from_slice(&tonemapping_lut_entries); if cfg!(any(not(feature = "webgl"), not(target_arch = "wasm32"))) || (cfg!(all(feature = "webgl", target_arch = "wasm32")) && !multisampled) { entries.extend_from_slice(&prepass::get_bind_group_layout_entries( - [16, 17, 18], + [17, 18, 19], multisampled, )); } @@ -586,8 +597,9 @@ bitflags::bitflags! { const MAY_DISCARD = (1 << 6); // Guards shader codepaths that may discard, allowing early depth tests in most cases // See: https://www.khronos.org/opengl/wiki/Early_Fragment_Test const ENVIRONMENT_MAP = (1 << 7); - const DEPTH_CLAMP_ORTHO = (1 << 8); - const TAA = (1 << 9); + const SCREEN_SPACE_AMBIENT_OCCLUSION = (1 << 8); + const DEPTH_CLAMP_ORTHO = (1 << 9); + const TAA = (1 << 10); const BLEND_RESERVED_BITS = Self::BLEND_MASK_BITS << Self::BLEND_SHIFT_BITS; // ← Bitmask reserving bits for the blend state const BLEND_OPAQUE = (0 << Self::BLEND_SHIFT_BITS); // ← Values are just sequential within the mask, and can range from 0 to 3 const BLEND_PREMULTIPLIED_ALPHA = (1 << Self::BLEND_SHIFT_BITS); // @@ -727,6 +739,10 @@ impl SpecializedMeshPipeline for MeshPipeline { bind_group_layout.push(self.mesh_layout.clone()); }; + if key.contains(MeshPipelineKey::SCREEN_SPACE_AMBIENT_OCCLUSION) { + shader_defs.push("SCREEN_SPACE_AMBIENT_OCCLUSION".into()); + } + let vertex_buffer_layout = layout.get_layout(&vertex_attributes)?; let (label, blend, depth_write_enabled); @@ -974,6 +990,7 @@ pub fn queue_mesh_view_bind_groups( Entity, &ViewShadowBindings, &ViewClusterBindings, + Option<&ScreenSpaceAmbientOcclusionTextures>, Option<&ViewPrepassTextures>, Option<&EnvironmentMapLight>, &Tonemapping, @@ -1003,11 +1020,17 @@ pub fn queue_mesh_view_bind_groups( entity, view_shadow_bindings, view_cluster_bindings, + ssao_textures, prepass_textures, environment_map, tonemapping, ) in &views { + let fallback_ssao = fallback_images + .image_for_samplecount(1) + .texture_view + .clone(); + let layout = if msaa.samples() > 1 { &mesh_pipeline.view_layout_multisampled } else { @@ -1063,18 +1086,26 @@ pub fn queue_mesh_view_bind_groups( binding: 10, resource: fog_binding.clone(), }, + BindGroupEntry { + binding: 11, + resource: BindingResource::TextureView( + ssao_textures + .map(|t| &t.screen_space_ambient_occlusion_texture.default_view) + .unwrap_or(&fallback_ssao), + ), + }, ]; let env_map = environment_map::get_bindings( environment_map, &images, &fallback_cubemap, - [11, 12, 13], + [12, 13, 14], ); entries.extend_from_slice(&env_map); let tonemapping_luts = - get_lut_bindings(&images, &tonemapping_luts, tonemapping, [14, 15]); + get_lut_bindings(&images, &tonemapping_luts, tonemapping, [15, 16]); entries.extend_from_slice(&tonemapping_luts); // When using WebGL, we can't have a depth texture with multisampling @@ -1086,7 +1117,7 @@ pub fn queue_mesh_view_bind_groups( &mut fallback_images, &mut fallback_depths, &msaa, - [16, 17, 18], + [17, 18, 19], )); } diff --git a/crates/bevy_pbr/src/render/mesh_view_bindings.wgsl b/crates/bevy_pbr/src/render/mesh_view_bindings.wgsl index 378cf8e290935..4c3e20902d84d 100644 --- a/crates/bevy_pbr/src/render/mesh_view_bindings.wgsl +++ b/crates/bevy_pbr/src/render/mesh_view_bindings.wgsl @@ -47,29 +47,32 @@ var globals: Globals; var fog: Fog; @group(0) @binding(11) -var environment_map_diffuse: texture_cube; +var screen_space_ambient_occlusion_texture: texture_2d; + @group(0) @binding(12) -var environment_map_specular: texture_cube; +var environment_map_diffuse: texture_cube; @group(0) @binding(13) +var environment_map_specular: texture_cube; +@group(0) @binding(14) var environment_map_sampler: sampler; -@group(0) @binding(14) -var dt_lut_texture: texture_3d; @group(0) @binding(15) +var dt_lut_texture: texture_3d; +@group(0) @binding(16) var dt_lut_sampler: sampler; #ifdef MULTISAMPLED -@group(0) @binding(16) -var depth_prepass_texture: texture_depth_multisampled_2d; @group(0) @binding(17) -var normal_prepass_texture: texture_multisampled_2d; +var depth_prepass_texture: texture_depth_multisampled_2d; @group(0) @binding(18) +var normal_prepass_texture: texture_multisampled_2d; +@group(0) @binding(19) var motion_vector_prepass_texture: texture_multisampled_2d; #else -@group(0) @binding(16) -var depth_prepass_texture: texture_depth_2d; @group(0) @binding(17) -var normal_prepass_texture: texture_2d; +var depth_prepass_texture: texture_depth_2d; @group(0) @binding(18) +var normal_prepass_texture: texture_2d; +@group(0) @binding(19) var motion_vector_prepass_texture: texture_2d; #endif diff --git a/crates/bevy_pbr/src/render/pbr.wgsl b/crates/bevy_pbr/src/render/pbr.wgsl index 7043c4e92abe7..e342823b8f7c4 100644 --- a/crates/bevy_pbr/src/render/pbr.wgsl +++ b/crates/bevy_pbr/src/render/pbr.wgsl @@ -13,6 +13,10 @@ #import bevy_pbr::prepass_utils +#ifdef SCREEN_SPACE_AMBIENT_OCCLUSION +#import bevy_pbr::gtao_utils +#endif + struct FragmentInput { @builtin(front_facing) is_front: bool, @builtin(position) frag_coord: vec4, @@ -88,12 +92,20 @@ fn fragment(in: FragmentInput) -> @location(0) vec4 { pbr_input.material.metallic = metallic; pbr_input.material.perceptual_roughness = perceptual_roughness; - var occlusion: f32 = 1.0; + // TODO: Split into diffuse/specular occlusion? + var occlusion: vec3 = vec3(1.0); #ifdef VERTEX_UVS if ((material.flags & STANDARD_MATERIAL_FLAGS_OCCLUSION_TEXTURE_BIT) != 0u) { - occlusion = textureSample(occlusion_texture, occlusion_sampler, uv).r; + occlusion = vec3(textureSample(occlusion_texture, occlusion_sampler, in.uv).r); } #endif +#ifdef SCREEN_SPACE_AMBIENT_OCCLUSION + let ssao = textureLoad(screen_space_ambient_occlusion_texture, vec2(in.frag_coord.xy), 0i).r; + let ssao_multibounce = gtao_multibounce(ssao, pbr_input.material.base_color.rgb); + occlusion = min(occlusion, ssao_multibounce); +#endif + pbr_input.occlusion = occlusion; + pbr_input.frag_coord = in.frag_coord; pbr_input.world_position = in.world_position; diff --git a/crates/bevy_pbr/src/render/pbr_ambient.wgsl b/crates/bevy_pbr/src/render/pbr_ambient.wgsl index 054f28617078a..c2e33adda9bc8 100644 --- a/crates/bevy_pbr/src/render/pbr_ambient.wgsl +++ b/crates/bevy_pbr/src/render/pbr_ambient.wgsl @@ -10,7 +10,7 @@ fn ambient_light( diffuse_color: vec3, specular_color: vec3, perceptual_roughness: f32, - occlusion: f32, + occlusion: vec3, ) -> vec3 { let diffuse_ambient = EnvBRDFApprox(diffuse_color, F_AB(1.0, NdotV)) * occlusion; let specular_ambient = EnvBRDFApprox(specular_color, F_AB(perceptual_roughness, NdotV)); diff --git a/crates/bevy_pbr/src/render/pbr_functions.wgsl b/crates/bevy_pbr/src/render/pbr_functions.wgsl index ee6ac3a79ce47..993efe4f1b95b 100644 --- a/crates/bevy_pbr/src/render/pbr_functions.wgsl +++ b/crates/bevy_pbr/src/render/pbr_functions.wgsl @@ -126,7 +126,7 @@ fn calculate_view( struct PbrInput { material: StandardMaterial, - occlusion: f32, + occlusion: vec3, frag_coord: vec4, world_position: vec4, // Normalized world normal used for shadow mapping as normal-mapping is not used for shadow @@ -146,7 +146,7 @@ fn pbr_input_new() -> PbrInput { var pbr_input: PbrInput; pbr_input.material = standard_material_new(); - pbr_input.occlusion = 1.0; + pbr_input.occlusion = vec3(1.0); pbr_input.frag_coord = vec4(0.0, 0.0, 0.0, 1.0); pbr_input.world_position = vec4(0.0, 0.0, 0.0, 1.0); diff --git a/crates/bevy_pbr/src/render/utils.wgsl b/crates/bevy_pbr/src/render/utils.wgsl index cb63273171479..6c8a87a5a43ef 100644 --- a/crates/bevy_pbr/src/render/utils.wgsl +++ b/crates/bevy_pbr/src/render/utils.wgsl @@ -1,6 +1,7 @@ #define_import_path bevy_pbr::utils const PI: f32 = 3.141592653589793; +const HALF_PI: f32 = 1.57079632679; const E: f32 = 2.718281828459045; fn hsv2rgb(hue: f32, saturation: f32, value: f32) -> vec3 { diff --git a/crates/bevy_pbr/src/ssao/gtao.wgsl b/crates/bevy_pbr/src/ssao/gtao.wgsl new file mode 100644 index 0000000000000..18aa8d5ac8a0d --- /dev/null +++ b/crates/bevy_pbr/src/ssao/gtao.wgsl @@ -0,0 +1,176 @@ +// Ground Truth-based Ambient Occlusion (GTAO) +// Paper: https://www.activision.com/cdn/research/Practical_Real_Time_Strategies_for_Accurate_Indirect_Occlusion_NEW%20VERSION_COLOR.pdf +// Presentation: https://blog.selfshadow.com/publications/s2016-shading-course/activision/s2016_pbs_activision_occlusion.pdf + +// Source code heavily based on XeGTAO v1.30 from Intel +// https://github.com/GameTechDev/XeGTAO/blob/0d177ce06bfa642f64d8af4de1197ad1bcb862d4/Source/Rendering/Shaders/XeGTAO.hlsli + +#import bevy_pbr::gtao_utils +#import bevy_pbr::utils +#import bevy_render::view +#import bevy_render::globals + +@group(0) @binding(0) var preprocessed_depth: texture_2d; +@group(0) @binding(1) var normals: texture_2d; +@group(0) @binding(2) var hilbert_index_lut: texture_2d; +@group(0) @binding(3) var ambient_occlusion: texture_storage_2d; +@group(0) @binding(4) var depth_differences: texture_storage_2d; +@group(0) @binding(5) var globals: Globals; +@group(1) @binding(0) var point_clamp_sampler: sampler; +@group(1) @binding(1) var view: View; + +fn load_noise(pixel_coordinates: vec2) -> vec2 { + var index = textureLoad(hilbert_index_lut, pixel_coordinates % 64, 0).r; + +#ifdef TEMPORAL_NOISE + index += 288u * (globals.frame_count % 64u); +#endif + + // R2 sequence - http://extremelearning.com.au/unreasonable-effectiveness-of-quasirandom-sequences + return fract(0.5 + f32(index) * vec2(0.75487766624669276005, 0.5698402909980532659114)); +} + +// Calculate differences in depth between neighbor pixels (later used by the spatial denoiser pass to preserve object edges) +fn calculate_neighboring_depth_differences(pixel_coordinates: vec2) -> f32 { + // Sample the pixel's depth and 4 depths around it + let uv = vec2(pixel_coordinates) / view.viewport.zw; + let depths_upper_left = textureGather(0, preprocessed_depth, point_clamp_sampler, uv); + let depths_bottom_right = textureGather(0, preprocessed_depth, point_clamp_sampler, uv, vec2(1i, 1i)); + let depth_center = depths_upper_left.y; + let depth_left = depths_upper_left.x; + let depth_top = depths_upper_left.z; + let depth_bottom = depths_bottom_right.x; + let depth_right = depths_bottom_right.z; + + // Calculate the depth differences (large differences represent object edges) + var edge_info = vec4(depth_left, depth_right, depth_top, depth_bottom) - depth_center; + let slope_left_right = (edge_info.y - edge_info.x) * 0.5; + let slope_top_bottom = (edge_info.w - edge_info.z) * 0.5; + let edge_info_slope_adjusted = edge_info + vec4(slope_left_right, -slope_left_right, slope_top_bottom, -slope_top_bottom); + edge_info = min(abs(edge_info), abs(edge_info_slope_adjusted)); + let bias = 0.25; // Using the bias and then saturating nudges the values a bit + let scale = depth_center * 0.011; // Weight the edges by their distance from the camera + edge_info = saturate((1.0 + bias) - edge_info / scale); // Apply the bias and scale, and invert edge_info so that small values become large, and vice versa + + // Pack the edge info into the texture + let edge_info_packed = vec4(mypack4x8unorm(edge_info), 0u, 0u, 0u); + textureStore(depth_differences, pixel_coordinates, edge_info_packed); + + return depth_center; +} + +// TODO: Remove this once https://github.com/gfx-rs/naga/pull/2353 lands +fn mypack4x8unorm(e: vec4) -> u32 { + return u32(clamp(e.x, 0.0, 1.0) * 255.0 + 0.5) | + u32(clamp(e.y, 0.0, 1.0) * 255.0 + 0.5) << 8u | + u32(clamp(e.z, 0.0, 1.0) * 255.0 + 0.5) << 16u | + u32(clamp(e.w, 0.0, 1.0) * 255.0 + 0.5) << 24u; +} + +fn load_normal_view_space(uv: vec2) -> vec3 { + var world_normal = textureSampleLevel(normals, point_clamp_sampler, uv, 0.0).xyz; + world_normal = (world_normal * 2.0) - 1.0; + let inverse_view = mat3x3( + view.inverse_view[0].xyz, + view.inverse_view[1].xyz, + view.inverse_view[2].xyz, + ); + return inverse_view * world_normal; +} + +fn reconstruct_view_space_position(depth: f32, uv: vec2) -> vec3 { + let clip_xy = vec2(uv.x * 2.0 - 1.0, 1.0 - 2.0 * uv.y); + let t = view.inverse_projection * vec4(clip_xy, depth, 1.0); + let view_xyz = t.xyz / t.w; + return view_xyz; +} + +fn load_and_reconstruct_view_space_position(uv: vec2, sample_mip_level: f32) -> vec3 { + let depth = textureSampleLevel(preprocessed_depth, point_clamp_sampler, uv, sample_mip_level).r; + return reconstruct_view_space_position(depth, uv); +} + +@compute +@workgroup_size(8, 8, 1) +fn gtao(@builtin(global_invocation_id) global_id: vec3) { + let slice_count = f32(#SLICE_COUNT); + let samples_per_slice_side = f32(#SAMPLES_PER_SLICE_SIDE); + let effect_radius = 0.5 * 1.457; + let falloff_range = 0.615 * effect_radius; + let falloff_from = effect_radius * (1.0 - 0.615); + let falloff_mul = -1.0 / falloff_range; + let falloff_add = falloff_from / falloff_range + 1.0; + + let pixel_coordinates = vec2(global_id.xy); + let uv = (vec2(pixel_coordinates) + 0.5) / view.viewport.zw; + + var pixel_depth = calculate_neighboring_depth_differences(pixel_coordinates); + pixel_depth += 0.00001; // Avoid depth precision issues + + let pixel_position = reconstruct_view_space_position(pixel_depth, uv); + let pixel_normal = load_normal_view_space(uv); + let view_vec = normalize(-pixel_position); + + let noise = load_noise(pixel_coordinates); + let sample_scale = (-0.5 * effect_radius * view.projection[0][0]) / pixel_position.z; + + var visibility = 0.0; + for (var slice_t = 0.0; slice_t < slice_count; slice_t += 1.0) { + let slice = slice_t + noise.x; + let phi = (PI / slice_count) * slice; + let omega = vec2(cos(phi), sin(phi)); + + let direction = vec3(omega.xy, 0.0); + let orthographic_direction = direction - (dot(direction, view_vec) * view_vec); + let axis = cross(direction, view_vec); + let projected_normal = pixel_normal - axis * dot(pixel_normal, axis); + let projected_normal_length = length(projected_normal); + + let sign_norm = sign(dot(orthographic_direction, projected_normal)); + let cos_norm = saturate(dot(projected_normal, view_vec) / projected_normal_length); + let n = sign_norm * fast_acos(cos_norm); + + let min_cos_horizon_1 = cos(n + HALF_PI); + let min_cos_horizon_2 = cos(n - HALF_PI); + var cos_horizon_1 = min_cos_horizon_1; + var cos_horizon_2 = min_cos_horizon_2; + let sample_mul = vec2(omega.x, -omega.y) * sample_scale; + for (var sample_t = 0.0; sample_t < samples_per_slice_side; sample_t += 1.0) { + var sample_noise = (slice_t + sample_t * samples_per_slice_side) * 0.6180339887498948482; + sample_noise = fract(noise.y + sample_noise); + + var s = (sample_t + sample_noise) / samples_per_slice_side; + s *= s; // https://github.com/GameTechDev/XeGTAO#sample-distribution + let sample = s * sample_mul; + + let sample_mip_level = clamp(log2(length(sample)) - 3.3, 0.0, 5.0); // https://github.com/GameTechDev/XeGTAO#memory-bandwidth-bottleneck + let sample_position_1 = load_and_reconstruct_view_space_position(uv + sample, sample_mip_level); + let sample_position_2 = load_and_reconstruct_view_space_position(uv - sample, sample_mip_level); + + let sample_difference_1 = sample_position_1 - pixel_position; + let sample_difference_2 = sample_position_2 - pixel_position; + let sample_distance_1 = length(sample_difference_1); + let sample_distance_2 = length(sample_difference_2); + var sample_cos_horizon_1 = dot(sample_difference_1 / sample_distance_1, view_vec); + var sample_cos_horizon_2 = dot(sample_difference_2 / sample_distance_2, view_vec); + + let weight_1 = saturate(sample_distance_1 * falloff_mul + falloff_add); + let weight_2 = saturate(sample_distance_2 * falloff_mul + falloff_add); + sample_cos_horizon_1 = mix(min_cos_horizon_1, sample_cos_horizon_1, weight_1); + sample_cos_horizon_2 = mix(min_cos_horizon_2, sample_cos_horizon_2, weight_2); + + cos_horizon_1 = max(cos_horizon_1, sample_cos_horizon_1); + cos_horizon_2 = max(cos_horizon_2, sample_cos_horizon_2); + } + + let horizon_1 = fast_acos(cos_horizon_1); + let horizon_2 = -fast_acos(cos_horizon_2); + let v1 = (cos_norm + 2.0 * horizon_1 * sin(n) - cos(2.0 * horizon_1 - n)) / 4.0; + let v2 = (cos_norm + 2.0 * horizon_2 * sin(n) - cos(2.0 * horizon_2 - n)) / 4.0; + visibility += projected_normal_length * (v1 + v2); + } + visibility /= slice_count; + visibility = clamp(visibility, 0.03, 1.0); + + textureStore(ambient_occlusion, pixel_coordinates, vec4(visibility, 0.0, 0.0, 0.0)); +} diff --git a/crates/bevy_pbr/src/ssao/gtao_utils.wgsl b/crates/bevy_pbr/src/ssao/gtao_utils.wgsl new file mode 100644 index 0000000000000..3ada70d749423 --- /dev/null +++ b/crates/bevy_pbr/src/ssao/gtao_utils.wgsl @@ -0,0 +1,22 @@ +#define_import_path bevy_pbr::gtao_utils + +// Approximates single-bounce ambient occlusion to multi-bounce ambient occlusion +// https://blog.selfshadow.com/publications/s2016-shading-course/activision/s2016_pbs_activision_occlusion.pdf#page=78 +fn gtao_multibounce(visibility: f32, base_color: vec3) -> vec3 { + let a = 2.0404 * base_color - 0.3324; + let b = -4.7951 * base_color + 0.6417; + let c = 2.7552 * base_color + 0.6903; + let x = vec3(visibility); + return max(x, ((x * a + b) * x + c) * x); +} + +fn fast_sqrt(x: f32) -> f32 { + return bitcast(0x1fbd1df5 + (bitcast(x) >> 1u)); +} + +fn fast_acos(in_x: f32) -> f32 { + let x = abs(in_x); + var res = -0.156583 * x + HALF_PI; + res *= fast_sqrt(1.0 - x); + return select(PI - res, res, in_x >= 0.0); +} diff --git a/crates/bevy_pbr/src/ssao/mod.rs b/crates/bevy_pbr/src/ssao/mod.rs new file mode 100644 index 0000000000000..276ee261a5e01 --- /dev/null +++ b/crates/bevy_pbr/src/ssao/mod.rs @@ -0,0 +1,989 @@ +use bevy_app::{App, Plugin}; +use bevy_asset::{load_internal_asset, HandleUntyped}; +use bevy_core_pipeline::{ + core_3d::CORE_3D, + prelude::Camera3d, + prepass::{DepthPrepass, NormalPrepass, ViewPrepassTextures}, +}; +use bevy_ecs::{ + prelude::{Bundle, Component, Entity}, + query::{QueryItem, With}, + reflect::ReflectComponent, + schedule::IntoSystemConfigs, + system::{Commands, Query, Res, ResMut, Resource}, + world::{FromWorld, World}, +}; +use bevy_reflect::{Reflect, TypeUuid}; +use bevy_render::{ + camera::{ExtractedCamera, TemporalJitter}, + extract_component::ExtractComponent, + globals::{GlobalsBuffer, GlobalsUniform}, + prelude::Camera, + render_graph::{NodeRunError, RenderGraphApp, RenderGraphContext, ViewNode, ViewNodeRunner}, + render_resource::{ + AddressMode, BindGroup, BindGroupDescriptor, BindGroupEntry, BindGroupLayout, + BindGroupLayoutDescriptor, BindGroupLayoutEntry, BindingResource, BindingType, + BufferBindingType, CachedComputePipelineId, ComputePassDescriptor, + ComputePipelineDescriptor, Extent3d, FilterMode, PipelineCache, Sampler, + SamplerBindingType, SamplerDescriptor, Shader, ShaderDefVal, ShaderStages, ShaderType, + SpecializedComputePipeline, SpecializedComputePipelines, StorageTextureAccess, + TextureDescriptor, TextureDimension, TextureFormat, TextureSampleType, TextureUsages, + TextureView, TextureViewDescriptor, TextureViewDimension, + }, + renderer::{RenderAdapter, RenderContext, RenderDevice, RenderQueue}, + texture::{CachedTexture, TextureCache}, + view::{Msaa, ViewUniform, ViewUniformOffset, ViewUniforms}, + Extract, ExtractSchedule, Render, RenderApp, RenderSet, +}; +use bevy_utils::{ + prelude::default, + tracing::{error, warn}, +}; +use std::mem; + +pub mod draw_3d_graph { + pub mod node { + /// Label for the screen space ambient occlusion render node. + pub const SCREEN_SPACE_AMBIENT_OCCLUSION: &str = "screen_space_ambient_occlusion"; + } +} + +const PREPROCESS_DEPTH_SHADER_HANDLE: HandleUntyped = + HandleUntyped::weak_from_u64(Shader::TYPE_UUID, 102258915420479); +const GTAO_SHADER_HANDLE: HandleUntyped = + HandleUntyped::weak_from_u64(Shader::TYPE_UUID, 253938746510568); +const SPATIAL_DENOISE_SHADER_HANDLE: HandleUntyped = + HandleUntyped::weak_from_u64(Shader::TYPE_UUID, 466162052558226); +const GTAO_UTILS_SHADER_HANDLE: HandleUntyped = + HandleUntyped::weak_from_u64(Shader::TYPE_UUID, 366465052568786); + +/// Plugin for screen space ambient occlusion. +pub struct ScreenSpaceAmbientOcclusionPlugin; + +impl Plugin for ScreenSpaceAmbientOcclusionPlugin { + fn build(&self, app: &mut App) { + load_internal_asset!( + app, + PREPROCESS_DEPTH_SHADER_HANDLE, + "preprocess_depth.wgsl", + Shader::from_wgsl + ); + load_internal_asset!(app, GTAO_SHADER_HANDLE, "gtao.wgsl", Shader::from_wgsl); + load_internal_asset!( + app, + SPATIAL_DENOISE_SHADER_HANDLE, + "spatial_denoise.wgsl", + Shader::from_wgsl + ); + load_internal_asset!( + app, + GTAO_UTILS_SHADER_HANDLE, + "gtao_utils.wgsl", + Shader::from_wgsl + ); + + app.register_type::(); + } + + fn finish(&self, app: &mut App) { + let Ok(render_app) = app.get_sub_app_mut(RenderApp) else { return }; + + if !render_app + .world + .resource::() + .get_texture_format_features(TextureFormat::R16Float) + .allowed_usages + .contains(TextureUsages::STORAGE_BINDING) + { + warn!("ScreenSpaceAmbientOcclusionPlugin not loaded. GPU lacks support: TextureFormat::R16Float does not support TextureUsages::STORAGE_BINDING."); + return; + } + + if render_app + .world + .resource::() + .limits() + .max_storage_textures_per_shader_stage + < 5 + { + warn!("ScreenSpaceAmbientOcclusionPlugin not loaded. GPU lacks support: Limits::max_storage_textures_per_shader_stage is less than 5."); + return; + } + + render_app + .init_resource::() + .init_resource::>() + .add_systems(ExtractSchedule, extract_ssao_settings) + .add_systems(Render, prepare_ssao_textures.in_set(RenderSet::Prepare)) + .add_systems(Render, prepare_ssao_pipelines.in_set(RenderSet::Prepare)) + .add_systems(Render, queue_ssao_bind_groups.in_set(RenderSet::Queue)) + .add_render_graph_node::>( + CORE_3D, + draw_3d_graph::node::SCREEN_SPACE_AMBIENT_OCCLUSION, + ) + .add_render_graph_edges( + CORE_3D, + &[ + // PREPASS -> SCREEN_SPACE_AMBIENT_OCCLUSION -> MAIN_PASS + bevy_core_pipeline::core_3d::graph::node::PREPASS, + draw_3d_graph::node::SCREEN_SPACE_AMBIENT_OCCLUSION, + bevy_core_pipeline::core_3d::graph::node::START_MAIN_PASS, + ], + ); + } +} + +/// Bundle to apply screen space ambient occlusion. +#[derive(Bundle, Default)] +pub struct ScreenSpaceAmbientOcclusionBundle { + pub settings: ScreenSpaceAmbientOcclusionSettings, + pub depth_prepass: DepthPrepass, + pub normal_prepass: NormalPrepass, +} + +/// Component to apply screen space ambient occlusion to a 3d camera. +/// +/// Screen space ambient occlusion (SSAO) approximates small-scale, +/// local occlusion of _indirect_ diffuse light between objects, based on what's visible on-screen. +/// SSAO does not apply to direct lighting, such as point or directional lights. +/// +/// This darkens creases, e.g. on staircases, and gives nice contact shadows +/// where objects meet, giving entities a more "grounded" feel. +/// +/// # Usage Notes +/// +/// Requires that you add [`ScreenSpaceAmbientOcclusionPlugin`] to your app, +/// and add the [`DepthPrepass`] and [`NormalPrepass`] components to your camera. +/// +/// It strongly recommended that you use SSAO in conjunction with +/// TAA ([`bevy_core_pipeline::experimental::taa::TemporalAntiAliasSettings`]). +/// Doing so greatly reduces SSAO noise. +/// +/// SSAO is not supported on `WebGL2`, and is not currently supported on `WebGPU` or `DirectX12`. +#[derive(Component, ExtractComponent, Reflect, PartialEq, Eq, Hash, Clone, Default)] +#[reflect(Component)] +pub struct ScreenSpaceAmbientOcclusionSettings { + pub quality_level: ScreenSpaceAmbientOcclusionQualityLevel, +} + +#[derive(Reflect, PartialEq, Eq, Hash, Clone, Copy, Default)] +pub enum ScreenSpaceAmbientOcclusionQualityLevel { + Low, + Medium, + #[default] + High, + Ultra, + Custom { + /// Higher slice count means less noise, but worse performance. + slice_count: u32, + /// Samples per slice side is also tweakable, but recommended to be left at 2 or 3. + samples_per_slice_side: u32, + }, +} + +impl ScreenSpaceAmbientOcclusionQualityLevel { + fn sample_counts(&self) -> (u32, u32) { + match self { + Self::Low => (1, 2), // 4 spp (1 * (2 * 2)), plus optional temporal samples + Self::Medium => (2, 2), // 8 spp (2 * (2 * 2)), plus optional temporal samples + Self::High => (3, 3), // 18 spp (3 * (3 * 2)), plus optional temporal samples + Self::Ultra => (9, 3), // 54 spp (9 * (3 * 2)), plus optional temporal samples + Self::Custom { + slice_count: slices, + samples_per_slice_side, + } => (*slices, *samples_per_slice_side), + } + } +} + +#[derive(Default)] +struct SsaoNode {} + +impl ViewNode for SsaoNode { + type ViewQuery = ( + &'static ExtractedCamera, + &'static SsaoPipelineId, + &'static SsaoBindGroups, + &'static ViewUniformOffset, + ); + + fn run( + &self, + _graph: &mut RenderGraphContext, + render_context: &mut RenderContext, + (camera, pipeline_id, bind_groups, view_uniform_offset): QueryItem, + world: &World, + ) -> Result<(), NodeRunError> { + let pipelines = world.resource::(); + let pipeline_cache = world.resource::(); + let ( + Some(camera_size), + Some(preprocess_depth_pipeline), + Some(spatial_denoise_pipeline), + Some(gtao_pipeline), + ) = ( + camera.physical_viewport_size, + pipeline_cache.get_compute_pipeline(pipelines.preprocess_depth_pipeline), + pipeline_cache.get_compute_pipeline(pipelines.spatial_denoise_pipeline), + pipeline_cache.get_compute_pipeline(pipeline_id.0), + ) else { + return Ok(()); + }; + + render_context.command_encoder().push_debug_group("ssao"); + + { + let mut preprocess_depth_pass = + render_context + .command_encoder() + .begin_compute_pass(&ComputePassDescriptor { + label: Some("ssao_preprocess_depth_pass"), + }); + preprocess_depth_pass.set_pipeline(preprocess_depth_pipeline); + preprocess_depth_pass.set_bind_group(0, &bind_groups.preprocess_depth_bind_group, &[]); + preprocess_depth_pass.set_bind_group( + 1, + &bind_groups.common_bind_group, + &[view_uniform_offset.offset], + ); + preprocess_depth_pass.dispatch_workgroups( + div_ceil(camera_size.x, 16), + div_ceil(camera_size.y, 16), + 1, + ); + } + + { + let mut gtao_pass = + render_context + .command_encoder() + .begin_compute_pass(&ComputePassDescriptor { + label: Some("ssao_gtao_pass"), + }); + gtao_pass.set_pipeline(gtao_pipeline); + gtao_pass.set_bind_group(0, &bind_groups.gtao_bind_group, &[]); + gtao_pass.set_bind_group( + 1, + &bind_groups.common_bind_group, + &[view_uniform_offset.offset], + ); + gtao_pass.dispatch_workgroups( + div_ceil(camera_size.x, 8), + div_ceil(camera_size.y, 8), + 1, + ); + } + + { + let mut spatial_denoise_pass = + render_context + .command_encoder() + .begin_compute_pass(&ComputePassDescriptor { + label: Some("ssao_spatial_denoise_pass"), + }); + spatial_denoise_pass.set_pipeline(spatial_denoise_pipeline); + spatial_denoise_pass.set_bind_group(0, &bind_groups.spatial_denoise_bind_group, &[]); + spatial_denoise_pass.set_bind_group( + 1, + &bind_groups.common_bind_group, + &[view_uniform_offset.offset], + ); + spatial_denoise_pass.dispatch_workgroups( + div_ceil(camera_size.x, 8), + div_ceil(camera_size.y, 8), + 1, + ); + } + + render_context.command_encoder().pop_debug_group(); + Ok(()) + } +} + +#[derive(Resource)] +struct SsaoPipelines { + preprocess_depth_pipeline: CachedComputePipelineId, + spatial_denoise_pipeline: CachedComputePipelineId, + + common_bind_group_layout: BindGroupLayout, + preprocess_depth_bind_group_layout: BindGroupLayout, + gtao_bind_group_layout: BindGroupLayout, + spatial_denoise_bind_group_layout: BindGroupLayout, + + hilbert_index_lut: TextureView, + point_clamp_sampler: Sampler, +} + +impl FromWorld for SsaoPipelines { + fn from_world(world: &mut World) -> Self { + let render_device = world.resource::(); + let render_queue = world.resource::(); + let pipeline_cache = world.resource::(); + + let hilbert_index_lut = render_device + .create_texture_with_data( + render_queue, + &(TextureDescriptor { + label: Some("ssao_hilbert_index_lut"), + size: Extent3d { + width: HILBERT_WIDTH as u32, + height: HILBERT_WIDTH as u32, + depth_or_array_layers: 1, + }, + mip_level_count: 1, + sample_count: 1, + dimension: TextureDimension::D2, + format: TextureFormat::R16Uint, + usage: TextureUsages::TEXTURE_BINDING, + view_formats: &[], + }), + bytemuck::cast_slice(&generate_hilbert_index_lut()), + ) + .create_view(&TextureViewDescriptor::default()); + + let point_clamp_sampler = render_device.create_sampler(&SamplerDescriptor { + min_filter: FilterMode::Nearest, + mag_filter: FilterMode::Nearest, + mipmap_filter: FilterMode::Nearest, + address_mode_u: AddressMode::ClampToEdge, + address_mode_v: AddressMode::ClampToEdge, + ..Default::default() + }); + + let common_bind_group_layout = + render_device.create_bind_group_layout(&BindGroupLayoutDescriptor { + label: Some("ssao_common_bind_group_layout"), + entries: &[ + BindGroupLayoutEntry { + binding: 0, + visibility: ShaderStages::COMPUTE, + ty: BindingType::Sampler(SamplerBindingType::NonFiltering), + count: None, + }, + BindGroupLayoutEntry { + binding: 1, + visibility: ShaderStages::COMPUTE, + ty: BindingType::Buffer { + ty: BufferBindingType::Uniform, + has_dynamic_offset: true, + min_binding_size: Some(ViewUniform::min_size()), + }, + count: None, + }, + ], + }); + + let mip_texture_entry = BindGroupLayoutEntry { + binding: 1, + visibility: ShaderStages::COMPUTE, + ty: BindingType::StorageTexture { + access: StorageTextureAccess::WriteOnly, + format: TextureFormat::R16Float, + view_dimension: TextureViewDimension::D2, + }, + count: None, + }; + let preprocess_depth_bind_group_layout = + render_device.create_bind_group_layout(&BindGroupLayoutDescriptor { + label: Some("ssao_preprocess_depth_bind_group_layout"), + entries: &[ + BindGroupLayoutEntry { + binding: 0, + visibility: ShaderStages::COMPUTE, + ty: BindingType::Texture { + sample_type: TextureSampleType::Depth, + view_dimension: TextureViewDimension::D2, + multisampled: false, + }, + count: None, + }, + mip_texture_entry, + BindGroupLayoutEntry { + binding: 2, + ..mip_texture_entry + }, + BindGroupLayoutEntry { + binding: 3, + ..mip_texture_entry + }, + BindGroupLayoutEntry { + binding: 4, + ..mip_texture_entry + }, + BindGroupLayoutEntry { + binding: 5, + ..mip_texture_entry + }, + ], + }); + + let gtao_bind_group_layout = + render_device.create_bind_group_layout(&BindGroupLayoutDescriptor { + label: Some("ssao_gtao_bind_group_layout"), + entries: &[ + BindGroupLayoutEntry { + binding: 0, + visibility: ShaderStages::COMPUTE, + ty: BindingType::Texture { + sample_type: TextureSampleType::Float { filterable: false }, + view_dimension: TextureViewDimension::D2, + multisampled: false, + }, + count: None, + }, + BindGroupLayoutEntry { + binding: 1, + visibility: ShaderStages::COMPUTE, + ty: BindingType::Texture { + sample_type: TextureSampleType::Float { filterable: false }, + view_dimension: TextureViewDimension::D2, + multisampled: false, + }, + count: None, + }, + BindGroupLayoutEntry { + binding: 2, + visibility: ShaderStages::COMPUTE, + ty: BindingType::Texture { + sample_type: TextureSampleType::Uint, + view_dimension: TextureViewDimension::D2, + multisampled: false, + }, + count: None, + }, + BindGroupLayoutEntry { + binding: 3, + visibility: ShaderStages::COMPUTE, + ty: BindingType::StorageTexture { + access: StorageTextureAccess::WriteOnly, + format: TextureFormat::R16Float, + view_dimension: TextureViewDimension::D2, + }, + count: None, + }, + BindGroupLayoutEntry { + binding: 4, + visibility: ShaderStages::COMPUTE, + ty: BindingType::StorageTexture { + access: StorageTextureAccess::WriteOnly, + format: TextureFormat::R32Uint, + view_dimension: TextureViewDimension::D2, + }, + count: None, + }, + BindGroupLayoutEntry { + binding: 5, + visibility: ShaderStages::COMPUTE, + ty: BindingType::Buffer { + ty: BufferBindingType::Uniform, + has_dynamic_offset: false, + min_binding_size: Some(GlobalsUniform::min_size()), + }, + count: None, + }, + ], + }); + + let spatial_denoise_bind_group_layout = + render_device.create_bind_group_layout(&BindGroupLayoutDescriptor { + label: Some("ssao_spatial_denoise_bind_group_layout"), + entries: &[ + BindGroupLayoutEntry { + binding: 0, + visibility: ShaderStages::COMPUTE, + ty: BindingType::Texture { + sample_type: TextureSampleType::Float { filterable: false }, + view_dimension: TextureViewDimension::D2, + multisampled: false, + }, + count: None, + }, + BindGroupLayoutEntry { + binding: 1, + visibility: ShaderStages::COMPUTE, + ty: BindingType::Texture { + sample_type: TextureSampleType::Uint, + view_dimension: TextureViewDimension::D2, + multisampled: false, + }, + count: None, + }, + BindGroupLayoutEntry { + binding: 2, + visibility: ShaderStages::COMPUTE, + ty: BindingType::StorageTexture { + access: StorageTextureAccess::WriteOnly, + format: TextureFormat::R16Float, + view_dimension: TextureViewDimension::D2, + }, + count: None, + }, + ], + }); + + let preprocess_depth_pipeline = + pipeline_cache.queue_compute_pipeline(ComputePipelineDescriptor { + label: Some("ssao_preprocess_depth_pipeline".into()), + layout: vec![ + preprocess_depth_bind_group_layout.clone(), + common_bind_group_layout.clone(), + ], + push_constant_ranges: vec![], + shader: PREPROCESS_DEPTH_SHADER_HANDLE.typed(), + shader_defs: Vec::new(), + entry_point: "preprocess_depth".into(), + }); + + let spatial_denoise_pipeline = + pipeline_cache.queue_compute_pipeline(ComputePipelineDescriptor { + label: Some("ssao_spatial_denoise_pipeline".into()), + layout: vec![ + spatial_denoise_bind_group_layout.clone(), + common_bind_group_layout.clone(), + ], + push_constant_ranges: vec![], + shader: SPATIAL_DENOISE_SHADER_HANDLE.typed(), + shader_defs: Vec::new(), + entry_point: "spatial_denoise".into(), + }); + + Self { + preprocess_depth_pipeline, + spatial_denoise_pipeline, + + common_bind_group_layout, + preprocess_depth_bind_group_layout, + gtao_bind_group_layout, + spatial_denoise_bind_group_layout, + + hilbert_index_lut, + point_clamp_sampler, + } + } +} + +#[derive(PartialEq, Eq, Hash, Clone)] +struct SsaoPipelineKey { + ssao_settings: ScreenSpaceAmbientOcclusionSettings, + temporal_noise: bool, +} + +impl SpecializedComputePipeline for SsaoPipelines { + type Key = SsaoPipelineKey; + + fn specialize(&self, key: Self::Key) -> ComputePipelineDescriptor { + let (slice_count, samples_per_slice_side) = key.ssao_settings.quality_level.sample_counts(); + + let mut shader_defs = vec![ + ShaderDefVal::Int("SLICE_COUNT".to_string(), slice_count as i32), + ShaderDefVal::Int( + "SAMPLES_PER_SLICE_SIDE".to_string(), + samples_per_slice_side as i32, + ), + ]; + + if key.temporal_noise { + shader_defs.push("TEMPORAL_NOISE".into()); + } + + ComputePipelineDescriptor { + label: Some("ssao_gtao_pipeline".into()), + layout: vec![ + self.gtao_bind_group_layout.clone(), + self.common_bind_group_layout.clone(), + ], + push_constant_ranges: vec![], + shader: GTAO_SHADER_HANDLE.typed(), + shader_defs, + entry_point: "gtao".into(), + } + } +} + +fn extract_ssao_settings( + mut commands: Commands, + cameras: Extract< + Query< + (Entity, &Camera, &ScreenSpaceAmbientOcclusionSettings), + (With, With, With), + >, + >, + msaa: Extract>, +) { + for (entity, camera, ssao_settings) in &cameras { + if **msaa != Msaa::Off { + error!( + "SSAO is being used which requires Msaa::Off, but Msaa is currently set to Msaa::{:?}", + **msaa + ); + return; + } + + if camera.is_active { + commands.get_or_spawn(entity).insert(ssao_settings.clone()); + } + } +} + +#[derive(Component)] +pub struct ScreenSpaceAmbientOcclusionTextures { + preprocessed_depth_texture: CachedTexture, + ssao_noisy_texture: CachedTexture, // Pre-spatially denoised texture + pub screen_space_ambient_occlusion_texture: CachedTexture, // Spatially denoised texture + depth_differences_texture: CachedTexture, +} + +fn prepare_ssao_textures( + mut commands: Commands, + mut texture_cache: ResMut, + render_device: Res, + views: Query<(Entity, &ExtractedCamera), With>, +) { + for (entity, camera) in &views { + let Some(physical_viewport_size) = camera.physical_viewport_size else { continue }; + let size = Extent3d { + width: physical_viewport_size.x, + height: physical_viewport_size.y, + depth_or_array_layers: 1, + }; + + let preprocessed_depth_texture = texture_cache.get( + &render_device, + TextureDescriptor { + label: Some("ssao_preprocessed_depth_texture"), + size, + mip_level_count: 5, + sample_count: 1, + dimension: TextureDimension::D2, + format: TextureFormat::R16Float, + usage: TextureUsages::STORAGE_BINDING | TextureUsages::TEXTURE_BINDING, + view_formats: &[], + }, + ); + + let ssao_noisy_texture = texture_cache.get( + &render_device, + TextureDescriptor { + label: Some("ssao_noisy_texture"), + size, + mip_level_count: 1, + sample_count: 1, + dimension: TextureDimension::D2, + format: TextureFormat::R16Float, + usage: TextureUsages::STORAGE_BINDING | TextureUsages::TEXTURE_BINDING, + view_formats: &[], + }, + ); + + let ssao_texture = texture_cache.get( + &render_device, + TextureDescriptor { + label: Some("ssao_texture"), + size, + mip_level_count: 1, + sample_count: 1, + dimension: TextureDimension::D2, + format: TextureFormat::R16Float, + usage: TextureUsages::STORAGE_BINDING | TextureUsages::TEXTURE_BINDING, + view_formats: &[], + }, + ); + + let depth_differences_texture = texture_cache.get( + &render_device, + TextureDescriptor { + label: Some("ssao_depth_differences_texture"), + size, + mip_level_count: 1, + sample_count: 1, + dimension: TextureDimension::D2, + format: TextureFormat::R32Uint, + usage: TextureUsages::STORAGE_BINDING | TextureUsages::TEXTURE_BINDING, + view_formats: &[], + }, + ); + + commands + .entity(entity) + .insert(ScreenSpaceAmbientOcclusionTextures { + preprocessed_depth_texture, + ssao_noisy_texture, + screen_space_ambient_occlusion_texture: ssao_texture, + depth_differences_texture, + }); + } +} + +#[derive(Component)] +struct SsaoPipelineId(CachedComputePipelineId); + +fn prepare_ssao_pipelines( + mut commands: Commands, + pipeline_cache: Res, + mut pipelines: ResMut>, + pipeline: Res, + views: Query<( + Entity, + &ScreenSpaceAmbientOcclusionSettings, + Option<&TemporalJitter>, + )>, +) { + for (entity, ssao_settings, temporal_jitter) in &views { + let pipeline_id = pipelines.specialize( + &pipeline_cache, + &pipeline, + SsaoPipelineKey { + ssao_settings: ssao_settings.clone(), + temporal_noise: temporal_jitter.is_some(), + }, + ); + + commands.entity(entity).insert(SsaoPipelineId(pipeline_id)); + } +} + +#[derive(Component)] +struct SsaoBindGroups { + common_bind_group: BindGroup, + preprocess_depth_bind_group: BindGroup, + gtao_bind_group: BindGroup, + spatial_denoise_bind_group: BindGroup, +} + +fn queue_ssao_bind_groups( + mut commands: Commands, + render_device: Res, + pipelines: Res, + view_uniforms: Res, + global_uniforms: Res, + views: Query<( + Entity, + &ScreenSpaceAmbientOcclusionTextures, + &ViewPrepassTextures, + )>, +) { + let (Some(view_uniforms), Some(globals_uniforms)) = ( + view_uniforms.uniforms.binding(), + global_uniforms.buffer.binding(), + ) else { + return; + }; + + for (entity, ssao_textures, prepass_textures) in &views { + let common_bind_group = render_device.create_bind_group(&BindGroupDescriptor { + label: Some("ssao_common_bind_group"), + layout: &pipelines.common_bind_group_layout, + entries: &[ + BindGroupEntry { + binding: 0, + resource: BindingResource::Sampler(&pipelines.point_clamp_sampler), + }, + BindGroupEntry { + binding: 1, + resource: view_uniforms.clone(), + }, + ], + }); + + let preprocess_depth_mip_view_descriptor = TextureViewDescriptor { + format: Some(TextureFormat::R16Float), + dimension: Some(TextureViewDimension::D2), + mip_level_count: Some(1), + ..default() + }; + let preprocess_depth_bind_group = render_device.create_bind_group(&BindGroupDescriptor { + label: Some("ssao_preprocess_depth_bind_group"), + layout: &pipelines.preprocess_depth_bind_group_layout, + entries: &[ + BindGroupEntry { + binding: 0, + resource: BindingResource::TextureView( + &prepass_textures.depth.as_ref().unwrap().default_view, + ), + }, + BindGroupEntry { + binding: 1, + resource: BindingResource::TextureView( + &ssao_textures + .preprocessed_depth_texture + .texture + .create_view(&TextureViewDescriptor { + label: Some("ssao_preprocessed_depth_texture_mip_view_0"), + base_mip_level: 0, + ..preprocess_depth_mip_view_descriptor + }), + ), + }, + BindGroupEntry { + binding: 2, + resource: BindingResource::TextureView( + &ssao_textures + .preprocessed_depth_texture + .texture + .create_view(&TextureViewDescriptor { + label: Some("ssao_preprocessed_depth_texture_mip_view_1"), + base_mip_level: 1, + ..preprocess_depth_mip_view_descriptor + }), + ), + }, + BindGroupEntry { + binding: 3, + resource: BindingResource::TextureView( + &ssao_textures + .preprocessed_depth_texture + .texture + .create_view(&TextureViewDescriptor { + label: Some("ssao_preprocessed_depth_texture_mip_view_2"), + base_mip_level: 2, + ..preprocess_depth_mip_view_descriptor + }), + ), + }, + BindGroupEntry { + binding: 4, + resource: BindingResource::TextureView( + &ssao_textures + .preprocessed_depth_texture + .texture + .create_view(&TextureViewDescriptor { + label: Some("ssao_preprocessed_depth_texture_mip_view_3"), + base_mip_level: 3, + ..preprocess_depth_mip_view_descriptor + }), + ), + }, + BindGroupEntry { + binding: 5, + resource: BindingResource::TextureView( + &ssao_textures + .preprocessed_depth_texture + .texture + .create_view(&TextureViewDescriptor { + label: Some("ssao_preprocessed_depth_texture_mip_view_4"), + base_mip_level: 4, + ..preprocess_depth_mip_view_descriptor + }), + ), + }, + ], + }); + + let gtao_bind_group = render_device.create_bind_group(&BindGroupDescriptor { + label: Some("ssao_gtao_bind_group"), + layout: &pipelines.gtao_bind_group_layout, + entries: &[ + BindGroupEntry { + binding: 0, + resource: BindingResource::TextureView( + &ssao_textures.preprocessed_depth_texture.default_view, + ), + }, + BindGroupEntry { + binding: 1, + resource: BindingResource::TextureView( + &prepass_textures.normal.as_ref().unwrap().default_view, + ), + }, + BindGroupEntry { + binding: 2, + resource: BindingResource::TextureView(&pipelines.hilbert_index_lut), + }, + BindGroupEntry { + binding: 3, + resource: BindingResource::TextureView( + &ssao_textures.ssao_noisy_texture.default_view, + ), + }, + BindGroupEntry { + binding: 4, + resource: BindingResource::TextureView( + &ssao_textures.depth_differences_texture.default_view, + ), + }, + BindGroupEntry { + binding: 5, + resource: globals_uniforms.clone(), + }, + ], + }); + + let spatial_denoise_bind_group = render_device.create_bind_group(&BindGroupDescriptor { + label: Some("ssao_spatial_denoise_bind_group"), + layout: &pipelines.spatial_denoise_bind_group_layout, + entries: &[ + BindGroupEntry { + binding: 0, + resource: BindingResource::TextureView( + &ssao_textures.ssao_noisy_texture.default_view, + ), + }, + BindGroupEntry { + binding: 1, + resource: BindingResource::TextureView( + &ssao_textures.depth_differences_texture.default_view, + ), + }, + BindGroupEntry { + binding: 2, + resource: BindingResource::TextureView( + &ssao_textures + .screen_space_ambient_occlusion_texture + .default_view, + ), + }, + ], + }); + + commands.entity(entity).insert(SsaoBindGroups { + common_bind_group, + preprocess_depth_bind_group, + gtao_bind_group, + spatial_denoise_bind_group, + }); + } +} + +#[allow(clippy::needless_range_loop)] +fn generate_hilbert_index_lut() -> [[u16; 64]; 64] { + let mut t = [[0; 64]; 64]; + + for x in 0..64 { + for y in 0..64 { + t[x][y] = hilbert_index(x as u16, y as u16); + } + } + + t +} + +// https://www.shadertoy.com/view/3tB3z3 +const HILBERT_WIDTH: u16 = 64; +fn hilbert_index(mut x: u16, mut y: u16) -> u16 { + let mut index = 0; + + let mut level: u16 = HILBERT_WIDTH / 2; + while level > 0 { + let region_x = (x & level > 0) as u16; + let region_y = (y & level > 0) as u16; + index += level * level * ((3 * region_x) ^ region_y); + + if region_y == 0 { + if region_x == 1 { + x = HILBERT_WIDTH - 1 - x; + y = HILBERT_WIDTH - 1 - y; + } + + mem::swap(&mut x, &mut y); + } + + level /= 2; + } + + index +} + +/// Divide `numerator` by `denominator`, rounded up to the nearest multiple of `denominator`. +fn div_ceil(numerator: u32, denominator: u32) -> u32 { + (numerator + denominator - 1) / denominator +} diff --git a/crates/bevy_pbr/src/ssao/preprocess_depth.wgsl b/crates/bevy_pbr/src/ssao/preprocess_depth.wgsl new file mode 100644 index 0000000000000..7ddcbb01265bf --- /dev/null +++ b/crates/bevy_pbr/src/ssao/preprocess_depth.wgsl @@ -0,0 +1,101 @@ +// Inputs a depth texture and outputs a MIP-chain of depths. +// +// Because SSAO's performance is bound by texture reads, this increases +// performance over using the full resolution depth for every sample. + +// Reference: https://research.nvidia.com/sites/default/files/pubs/2012-06_Scalable-Ambient-Obscurance/McGuire12SAO.pdf, section 2.2 + +#import bevy_render::view + +@group(0) @binding(0) var input_depth: texture_depth_2d; +@group(0) @binding(1) var preprocessed_depth_mip0: texture_storage_2d; +@group(0) @binding(2) var preprocessed_depth_mip1: texture_storage_2d; +@group(0) @binding(3) var preprocessed_depth_mip2: texture_storage_2d; +@group(0) @binding(4) var preprocessed_depth_mip3: texture_storage_2d; +@group(0) @binding(5) var preprocessed_depth_mip4: texture_storage_2d; +@group(1) @binding(0) var point_clamp_sampler: sampler; +@group(1) @binding(1) var view: View; + + +// Using 4 depths from the previous MIP, compute a weighted average for the depth of the current MIP +fn weighted_average(depth0: f32, depth1: f32, depth2: f32, depth3: f32) -> f32 { + let depth_range_scale_factor = 0.75; + let effect_radius = depth_range_scale_factor * 0.5 * 1.457; + let falloff_range = 0.615 * effect_radius; + let falloff_from = effect_radius * (1.0 - 0.615); + let falloff_mul = -1.0 / falloff_range; + let falloff_add = falloff_from / falloff_range + 1.0; + + let min_depth = min(min(depth0, depth1), min(depth2, depth3)); + let weight0 = saturate((depth0 - min_depth) * falloff_mul + falloff_add); + let weight1 = saturate((depth1 - min_depth) * falloff_mul + falloff_add); + let weight2 = saturate((depth2 - min_depth) * falloff_mul + falloff_add); + let weight3 = saturate((depth3 - min_depth) * falloff_mul + falloff_add); + let weight_total = weight0 + weight1 + weight2 + weight3; + + return ((weight0 * depth0) + (weight1 * depth1) + (weight2 * depth2) + (weight3 * depth3)) / weight_total; +} + +// Used to share the depths from the previous MIP level between all invocations in a workgroup +var previous_mip_depth: array, 8>; + +@compute +@workgroup_size(8, 8, 1) +fn preprocess_depth(@builtin(global_invocation_id) global_id: vec3, @builtin(local_invocation_id) local_id: vec3) { + let base_coordinates = vec2(global_id.xy); + + // MIP 0 - Copy 4 texels from the input depth (per invocation, 8x8 invocations per workgroup) + let pixel_coordinates0 = base_coordinates * 2i; + let pixel_coordinates1 = pixel_coordinates0 + vec2(1i, 0i); + let pixel_coordinates2 = pixel_coordinates0 + vec2(0i, 1i); + let pixel_coordinates3 = pixel_coordinates0 + vec2(1i, 1i); + let depths_uv = vec2(pixel_coordinates0) / view.viewport.zw; + let depths = textureGather(0, input_depth, point_clamp_sampler, depths_uv, vec2(1i, 1i)); + textureStore(preprocessed_depth_mip0, pixel_coordinates0, vec4(depths.w, 0.0, 0.0, 0.0)); + textureStore(preprocessed_depth_mip0, pixel_coordinates1, vec4(depths.z, 0.0, 0.0, 0.0)); + textureStore(preprocessed_depth_mip0, pixel_coordinates2, vec4(depths.x, 0.0, 0.0, 0.0)); + textureStore(preprocessed_depth_mip0, pixel_coordinates3, vec4(depths.y, 0.0, 0.0, 0.0)); + + // MIP 1 - Weighted average of MIP 0's depth values (per invocation, 8x8 invocations per workgroup) + let depth_mip1 = weighted_average(depths.w, depths.z, depths.x, depths.y); + textureStore(preprocessed_depth_mip1, base_coordinates, vec4(depth_mip1, 0.0, 0.0, 0.0)); + previous_mip_depth[local_id.x][local_id.y] = depth_mip1; + + workgroupBarrier(); + + // MIP 2 - Weighted average of MIP 1's depth values (per invocation, 4x4 invocations per workgroup) + if all(local_id.xy % vec2(2u) == vec2(0u)) { + let depth0 = previous_mip_depth[local_id.x + 0u][local_id.y + 0u]; + let depth1 = previous_mip_depth[local_id.x + 1u][local_id.y + 0u]; + let depth2 = previous_mip_depth[local_id.x + 0u][local_id.y + 1u]; + let depth3 = previous_mip_depth[local_id.x + 1u][local_id.y + 1u]; + let depth_mip2 = weighted_average(depth0, depth1, depth2, depth3); + textureStore(preprocessed_depth_mip2, base_coordinates / 2i, vec4(depth_mip2, 0.0, 0.0, 0.0)); + previous_mip_depth[local_id.x][local_id.y] = depth_mip2; + } + + workgroupBarrier(); + + // MIP 3 - Weighted average of MIP 2's depth values (per invocation, 2x2 invocations per workgroup) + if all(local_id.xy % vec2(4u) == vec2(0u)) { + let depth0 = previous_mip_depth[local_id.x + 0u][local_id.y + 0u]; + let depth1 = previous_mip_depth[local_id.x + 2u][local_id.y + 0u]; + let depth2 = previous_mip_depth[local_id.x + 0u][local_id.y + 2u]; + let depth3 = previous_mip_depth[local_id.x + 2u][local_id.y + 2u]; + let depth_mip3 = weighted_average(depth0, depth1, depth2, depth3); + textureStore(preprocessed_depth_mip3, base_coordinates / 4i, vec4(depth_mip3, 0.0, 0.0, 0.0)); + previous_mip_depth[local_id.x][local_id.y] = depth_mip3; + } + + workgroupBarrier(); + + // MIP 4 - Weighted average of MIP 3's depth values (per invocation, 1 invocation per workgroup) + if all(local_id.xy % vec2(8u) == vec2(0u)) { + let depth0 = previous_mip_depth[local_id.x + 0u][local_id.y + 0u]; + let depth1 = previous_mip_depth[local_id.x + 4u][local_id.y + 0u]; + let depth2 = previous_mip_depth[local_id.x + 0u][local_id.y + 4u]; + let depth3 = previous_mip_depth[local_id.x + 4u][local_id.y + 4u]; + let depth_mip4 = weighted_average(depth0, depth1, depth2, depth3); + textureStore(preprocessed_depth_mip4, base_coordinates / 8i, vec4(depth_mip4, 0.0, 0.0, 0.0)); + } +} diff --git a/crates/bevy_pbr/src/ssao/spatial_denoise.wgsl b/crates/bevy_pbr/src/ssao/spatial_denoise.wgsl new file mode 100644 index 0000000000000..64ce2f8dcb98c --- /dev/null +++ b/crates/bevy_pbr/src/ssao/spatial_denoise.wgsl @@ -0,0 +1,92 @@ +// 3x3 bilaterial filter (edge-preserving blur) +// https://people.csail.mit.edu/sparis/bf_course/course_notes.pdf + +// Note: Does not use the Gaussian kernel part of a typical bilateral blur +// From the paper: "use the information gathered on a neighborhood of 4 × 4 using a bilateral filter for +// reconstruction, using _uniform_ convolution weights" + +// Note: The paper does a 4x4 (not quite centered) filter, offset by +/- 1 pixel every other frame +// XeGTAO does a 3x3 filter, on two pixels at a time per compute thread, applied twice +// We do a 3x3 filter, on 1 pixel per compute thread, applied once + +#import bevy_render::view + +@group(0) @binding(0) var ambient_occlusion_noisy: texture_2d; +@group(0) @binding(1) var depth_differences: texture_2d; +@group(0) @binding(2) var ambient_occlusion: texture_storage_2d; +@group(1) @binding(0) var point_clamp_sampler: sampler; +@group(1) @binding(1) var view: View; + +@compute +@workgroup_size(8, 8, 1) +fn spatial_denoise(@builtin(global_invocation_id) global_id: vec3) { + let pixel_coordinates = vec2(global_id.xy); + let uv = vec2(pixel_coordinates) / view.viewport.zw; + + let edges0 = textureGather(0, depth_differences, point_clamp_sampler, uv); + let edges1 = textureGather(0, depth_differences, point_clamp_sampler, uv, vec2(2i, 0i)); + let edges2 = textureGather(0, depth_differences, point_clamp_sampler, uv, vec2(1i, 2i)); + let visibility0 = textureGather(0, ambient_occlusion_noisy, point_clamp_sampler, uv); + let visibility1 = textureGather(0, ambient_occlusion_noisy, point_clamp_sampler, uv, vec2(2i, 0i)); + let visibility2 = textureGather(0, ambient_occlusion_noisy, point_clamp_sampler, uv, vec2(0i, 2i)); + let visibility3 = textureGather(0, ambient_occlusion_noisy, point_clamp_sampler, uv, vec2(2i, 2i)); + + let left_edges = myunpack4x8unorm(edges0.x); + let right_edges = myunpack4x8unorm(edges1.x); + let top_edges = myunpack4x8unorm(edges0.z); + let bottom_edges = myunpack4x8unorm(edges2.w); + var center_edges = myunpack4x8unorm(edges0.y); + center_edges *= vec4(left_edges.y, right_edges.x, top_edges.w, bottom_edges.z); + + let center_weight = 1.2; + let left_weight = center_edges.x; + let right_weight = center_edges.y; + let top_weight = center_edges.z; + let bottom_weight = center_edges.w; + let top_left_weight = 0.425 * (top_weight * top_edges.x + left_weight * left_edges.z); + let top_right_weight = 0.425 * (top_weight * top_edges.y + right_weight * right_edges.z); + let bottom_left_weight = 0.425 * (bottom_weight * bottom_edges.x + left_weight * left_edges.w); + let bottom_right_weight = 0.425 * (bottom_weight * bottom_edges.y + right_weight * right_edges.w); + + let center_visibility = visibility0.y; + let left_visibility = visibility0.x; + let right_visibility = visibility0.z; + let top_visibility = visibility1.x; + let bottom_visibility = visibility2.z; + let top_left_visibility = visibility0.w; + let top_right_visibility = visibility1.w; + let bottom_left_visibility = visibility2.w; + let bottom_right_visibility = visibility3.w; + + var sum = center_visibility; + sum += left_visibility * left_weight; + sum += right_visibility * right_weight; + sum += top_visibility * top_weight; + sum += bottom_visibility * bottom_weight; + sum += top_left_visibility * top_left_weight; + sum += top_right_visibility * top_right_weight; + sum += bottom_left_visibility * bottom_left_weight; + sum += bottom_right_visibility * bottom_right_weight; + + var sum_weight = center_weight; + sum_weight += left_weight; + sum_weight += right_weight; + sum_weight += top_weight; + sum_weight += bottom_weight; + sum_weight += top_left_weight; + sum_weight += top_right_weight; + sum_weight += bottom_left_weight; + sum_weight += bottom_right_weight; + + let denoised_visibility = sum / sum_weight; + + textureStore(ambient_occlusion, pixel_coordinates, vec4(denoised_visibility, 0.0, 0.0, 0.0)); +} + +// TODO: Remove this once https://github.com/gfx-rs/naga/pull/2353 lands in Bevy +fn myunpack4x8unorm(e: u32) -> vec4 { + return vec4(clamp(f32(e & 0xFFu) / 255.0, 0.0, 1.0), + clamp(f32((e >> 8u) & 0xFFu) / 255.0, 0.0, 1.0), + clamp(f32((e >> 16u) & 0xFFu) / 255.0, 0.0, 1.0), + clamp(f32((e >> 24u) & 0xFFu) / 255.0, 0.0, 1.0)); +} diff --git a/crates/bevy_render/src/view/mod.rs b/crates/bevy_render/src/view/mod.rs index 2a8b226050cb1..c94f42e8ac8be 100644 --- a/crates/bevy_render/src/view/mod.rs +++ b/crates/bevy_render/src/view/mod.rs @@ -87,7 +87,9 @@ impl Plugin for ViewPlugin { /// .insert_resource(Msaa::default()) /// .run(); /// ``` -#[derive(Resource, Default, Clone, Copy, ExtractResource, Reflect, PartialEq, PartialOrd)] +#[derive( + Resource, Default, Clone, Copy, ExtractResource, Reflect, PartialEq, PartialOrd, Debug, +)] #[reflect(Resource)] pub enum Msaa { Off = 1, diff --git a/examples/3d/ssao.rs b/examples/3d/ssao.rs new file mode 100644 index 0000000000000..c95b71f45d25e --- /dev/null +++ b/examples/3d/ssao.rs @@ -0,0 +1,199 @@ +//! A scene showcasing screen space ambient occlusion. + +use bevy::{ + core_pipeline::experimental::taa::{TemporalAntiAliasBundle, TemporalAntiAliasPlugin}, + pbr::{ + ScreenSpaceAmbientOcclusionBundle, ScreenSpaceAmbientOcclusionQualityLevel, + ScreenSpaceAmbientOcclusionSettings, + }, + prelude::*, + render::camera::TemporalJitter, +}; +use std::f32::consts::PI; + +fn main() { + App::new() + .insert_resource(AmbientLight { + brightness: 5.0, + ..default() + }) + .add_plugins(DefaultPlugins) + .add_plugin(TemporalAntiAliasPlugin) + .add_systems(Startup, setup) + .add_systems(Update, update) + .run(); +} + +fn setup( + mut commands: Commands, + mut meshes: ResMut>, + mut materials: ResMut>, + asset_server: Res, +) { + commands + .spawn(Camera3dBundle { + camera: Camera { + hdr: true, + ..default() + }, + transform: Transform::from_xyz(-2.0, 2.0, -2.0).looking_at(Vec3::ZERO, Vec3::Y), + ..default() + }) + .insert(ScreenSpaceAmbientOcclusionBundle::default()) + .insert(TemporalAntiAliasBundle::default()); + + let material = materials.add(StandardMaterial { + base_color: Color::rgb(0.5, 0.5, 0.5), + perceptual_roughness: 1.0, + reflectance: 0.0, + ..default() + }); + commands.spawn(PbrBundle { + mesh: meshes.add(Mesh::from(shape::Cube { size: 1.0 })), + material: material.clone(), + transform: Transform::from_xyz(0.0, 0.0, 1.0), + ..default() + }); + commands.spawn(PbrBundle { + mesh: meshes.add(Mesh::from(shape::Cube { size: 1.0 })), + material: material.clone(), + transform: Transform::from_xyz(0.0, -1.0, 0.0), + ..default() + }); + commands.spawn(PbrBundle { + mesh: meshes.add(Mesh::from(shape::Cube { size: 1.0 })), + material, + transform: Transform::from_xyz(1.0, 0.0, 0.0), + ..default() + }); + commands.spawn(( + PbrBundle { + mesh: meshes.add(Mesh::from(shape::UVSphere { + radius: 0.4, + sectors: 72, + stacks: 36, + })), + material: materials.add(StandardMaterial { + base_color: Color::rgb(0.4, 0.4, 0.4), + perceptual_roughness: 1.0, + reflectance: 0.0, + ..default() + }), + ..default() + }, + SphereMarker, + )); + + commands.spawn(DirectionalLightBundle { + directional_light: DirectionalLight { + shadows_enabled: true, + ..default() + }, + transform: Transform::from_rotation(Quat::from_euler( + EulerRot::ZYX, + 0.0, + PI * -0.15, + PI * -0.15, + )), + ..default() + }); + + commands.spawn( + TextBundle::from_section( + "", + TextStyle { + font: asset_server.load("fonts/FiraMono-Medium.ttf"), + font_size: 26.0, + color: Color::BLACK, + }, + ) + .with_style(Style { + position_type: PositionType::Absolute, + bottom: Val::Px(10.0), + left: Val::Px(10.0), + ..default() + }), + ); +} + +fn update( + camera: Query< + ( + Entity, + Option<&ScreenSpaceAmbientOcclusionSettings>, + Option<&TemporalJitter>, + ), + With, + >, + mut text: Query<&mut Text>, + mut sphere: Query<&mut Transform, With>, + mut commands: Commands, + keycode: Res>, + time: Res