From 5569a8cdb31a12f7ec4b9a825379400edde211c8 Mon Sep 17 00:00:00 2001 From: Aditya Sharma Date: Mon, 27 Jul 2026 21:02:20 -0700 Subject: [PATCH 1/4] perf: improve backdrop blur quality --- crates/gpui/src/scene.rs | 205 +++++++ crates/gpui/src/styled.rs | 7 +- crates/gpui/src/window.rs | 8 + crates/gpui_macos/src/metal_renderer.rs | 604 ++++++++++++-------- crates/gpui_macos/src/shaders.metal | 52 +- crates/gpui_wgpu/src/shaders.wgsl | 140 +++-- crates/gpui_wgpu/src/wgpu_renderer.rs | 565 ++++++++++++++++-- crates/gpui_windows/src/directx_renderer.rs | 162 ++++-- crates/gpui_windows/src/shaders.hlsl | 60 +- 9 files changed, 1421 insertions(+), 382 deletions(-) diff --git a/crates/gpui/src/scene.rs b/crates/gpui/src/scene.rs index 96967b5..92cf8bd 100644 --- a/crates/gpui/src/scene.rs +++ b/crates/gpui/src/scene.rs @@ -697,6 +697,138 @@ impl From for Primitive { } } +/// Internal blur-pyramid configuration shared by renderer backends. +#[doc(hidden)] +#[derive(Copy, Clone, Debug)] +pub struct BackdropBlurPlan { + /// Number of downsample and upsample passes. + pub passes: usize, + /// Upsample tent-filter distance in input texels. + pub sample_distance: f32, +} + +impl BackdropBlurPlan { + /// Largest pyramid supported by renderer backends. + pub const MAX_PASSES: usize = 6; + + const DOWNSAMPLE_VARIANCE: f32 = 0.75; + const MAX_SAMPLE_DISTANCE: f32 = 1.5; + const SUPPORT_SIGMAS: f32 = 3.0; + const SOLVER_STEPS: usize = 20; + + /// A plan that samples the unblurred backdrop. + pub const IDENTITY: Self = Self { + passes: 0, + sample_distance: 0.0, + }; + + /// Builds a bounded pyramid whose three-sigma support approximates `radius`. + /// + /// Positive radii below the pyramid's minimum support use that minimum. + pub fn for_radius(radius: f32, max_passes: usize) -> Self { + if !radius.is_finite() || radius <= 0.0 || max_passes == 0 { + return Self::IDENTITY; + } + + let max_passes = max_passes.min(Self::MAX_PASSES); + let sigma_squared = (radius / Self::SUPPORT_SIGMAS).powi(2); + for passes in 1..=max_passes { + if sigma_squared <= Self::variance(passes, 0.0) { + return Self { + passes, + sample_distance: 0.0, + }; + } + + if sigma_squared <= Self::variance(passes, Self::MAX_SAMPLE_DISTANCE) { + let mut lower = 0.0; + let mut upper = Self::MAX_SAMPLE_DISTANCE; + for _ in 0..Self::SOLVER_STEPS { + let sample_distance = (lower + upper) * 0.5; + if Self::variance(passes, sample_distance) < sigma_squared { + lower = sample_distance; + } else { + upper = sample_distance; + } + } + return Self { + passes, + sample_distance: (lower + upper) * 0.5, + }; + } + } + + Self { + passes: max_passes, + sample_distance: Self::MAX_SAMPLE_DISTANCE, + } + } + + /// Returns source padding required by the supported three-sigma kernel. + pub fn padding(radius: f32) -> f32 { + if radius.is_finite() && radius > 0.0 { + let minimum_support = Self::variance(1, 0.0).sqrt() * Self::SUPPORT_SIGMAS; + let maximum_support = Self::variance(Self::MAX_PASSES, Self::MAX_SAMPLE_DISTANCE) + .sqrt() + * Self::SUPPORT_SIGMAS; + let support = radius.clamp(minimum_support, maximum_support); + (support + 2.0).ceil() + } else { + 0.0 + } + } + + fn variance(passes: usize, sample_distance: f32) -> f32 { + if passes == 0 { + return 0.0; + } + + let four_to_passes = 4.0_f32.powi(passes as i32); + let level_scale_sum = (four_to_passes - 1.0) / 3.0; + (Self::DOWNSAMPLE_VARIANCE + Self::upsample_variance(sample_distance)) * level_scale_sum + } + + fn upsample_variance(sample_distance: f32) -> f32 { + Self::linear_upsample_second_moment(0.0) / 6.0 + + (Self::linear_upsample_second_moment(sample_distance) + + Self::linear_upsample_second_moment(-sample_distance)) + / 3.0 + + (Self::linear_upsample_second_moment(2.0 * sample_distance) + + Self::linear_upsample_second_moment(-2.0 * sample_distance)) + / 12.0 + } + + fn linear_upsample_second_moment(offset: f32) -> f32 { + // Bilinear reconstruction of one input texel is a radius-two tent on + // the output lattice. Measure around 0.5, the shared centroid of the + // symmetric 8-tap upsample kernel. + let center = 0.5 - 2.0 * offset; + let first_output = center.floor() as i32 - 2; + let mut weight_sum = 0.0; + let mut second_moment = 0.0; + for output in first_output..=first_output + 5 { + let weight = (1.0 - ((output as f32 - center).abs() / 2.0)).max(0.0); + weight_sum += weight; + second_moment += weight * (output as f32 - 0.5).powi(2); + } + second_moment / weight_sum + } + + #[cfg(test)] + fn sigma(self) -> f32 { + Self::variance(self.passes, self.sample_distance).sqrt() + } +} + +impl PartialEq for BackdropBlurPlan { + fn eq(&self, other: &Self) -> bool { + self.passes == other.passes + && self.sample_distance.to_bits() == other.sample_distance.to_bits() + } +} + +impl Eq for BackdropBlurPlan {} + #[derive(Debug, Clone)] #[repr(C)] #[expect(missing_docs)] @@ -1178,6 +1310,79 @@ mod tests { assert_eq!(offset_of!(BackdropBlur, pad2), 108); } + #[test] + fn backdrop_blur_plan_preserves_radius_across_scale_and_large_values() { + let cases = [ + (0.0, 0), + (1.0, 1), + (11.0, 1), + (12.0, 2), + (24.0, 2), + (25.0, 3), + (50.0, 3), + (51.0, 4), + (101.0, 4), + (102.0, 5), + (203.0, 5), + (204.0, 6), + (240.0, 6), + ]; + + for (radius, expected_passes) in cases { + let plan = BackdropBlurPlan::for_radius(radius, BackdropBlurPlan::MAX_PASSES); + assert_eq!(plan.passes, expected_passes, "radius {radius}"); + if radius >= 4.0 { + assert!( + (plan.sigma() - radius / 3.0).abs() < 0.001, + "radius {radius}" + ); + } + } + } + + #[test] + fn backdrop_blur_plan_clamps_positive_radii_to_minimum_support() { + let plan = BackdropBlurPlan::for_radius(1.0, BackdropBlurPlan::MAX_PASSES); + assert_eq!(plan.passes, 1); + assert!((plan.sigma() * 3.0 - 3.674_234_6).abs() < 0.001); + } + + #[test] + fn backdrop_blur_plan_uses_available_levels_without_invalid_shader_values() { + let plan = BackdropBlurPlan::for_radius(240.0, 4); + assert_eq!(plan.passes, 4); + assert!(plan.sample_distance.is_finite()); + assert_eq!(plan.sample_distance, 1.5); + + for radius in [f32::NEG_INFINITY, -1.0, 0.0, f32::INFINITY, f32::NAN] { + assert_eq!( + BackdropBlurPlan::for_radius(radius, BackdropBlurPlan::MAX_PASSES), + BackdropBlurPlan::IDENTITY + ); + } + assert_eq!( + BackdropBlurPlan::for_radius(24.0, 0), + BackdropBlurPlan::IDENTITY + ); + } + + #[test] + fn backdrop_blur_plan_accounts_for_bilinear_upsample_variance() { + assert!((BackdropBlurPlan::upsample_variance(0.0) - 0.75).abs() < 0.0001); + assert!((BackdropBlurPlan::upsample_variance(1.0) - 6.083_333).abs() < 0.0001); + assert!((BackdropBlurPlan::upsample_variance(1.5) - 12.75).abs() < 0.0001); + } + + #[test] + fn backdrop_blur_padding_tracks_kernel_support() { + assert_eq!(BackdropBlurPlan::padding(0.0), 0.0); + assert_eq!(BackdropBlurPlan::padding(1.0), 6.0); + assert_eq!(BackdropBlurPlan::padding(24.0), 26.0); + assert_eq!(BackdropBlurPlan::padding(120.0), 122.0); + assert!(BackdropBlurPlan::padding(1_000.0) < 500.0); + assert_eq!(BackdropBlurPlan::padding(f32::NAN), 0.0); + } + #[test] fn padded_bool32_has_initialized_u32_representation() { assert_eq!(size_of::(), size_of::()); diff --git a/crates/gpui/src/styled.rs b/crates/gpui/src/styled.rs index 5381b65..00c8ba1 100644 --- a/crates/gpui/src/styled.rs +++ b/crates/gpui/src/styled.rs @@ -759,7 +759,12 @@ pub trait Styled: Sized { self } - /// Sets the blur radius applied to content behind this element. + /// Sets the blur extent applied to content behind this element. + /// + /// The radius approximates the filter's three-sigma support. At render + /// scale, positive radii below the pyramid's roughly 3.7-device-pixel + /// minimum use that minimum. Non-positive and non-finite values disable + /// the effect. fn backdrop_blur(mut self, blur_radius: Pixels) -> Self { self.style().backdrop_blur = Some(blur_radius); self diff --git a/crates/gpui/src/window.rs b/crates/gpui/src/window.rs index 3dfd575..1aef6fd 100644 --- a/crates/gpui/src/window.rs +++ b/crates/gpui/src/window.rs @@ -3709,6 +3709,10 @@ impl Window { } /// Paint a backdrop blur region into the scene for the next frame. + /// + /// `blur_radius` approximates the filter's three-sigma support. At render + /// scale, positive radii below the pyramid's roughly 3.7-device-pixel + /// minimum use that minimum. pub fn paint_backdrop_blur( &mut self, bounds: Bounds, @@ -3716,6 +3720,10 @@ impl Window { blur_radius: Pixels, ) { self.invalidator.debug_assert_paint(); + if !blur_radius.0.is_finite() || blur_radius <= Pixels::ZERO { + return; + } + let scale_factor = self.scale_factor(); let content_mask = self.content_mask(); self.next_frame.scene.insert_primitive(BackdropBlur { diff --git a/crates/gpui_macos/src/metal_renderer.rs b/crates/gpui_macos/src/metal_renderer.rs index 3ae77fd..d04818e 100644 --- a/crates/gpui_macos/src/metal_renderer.rs +++ b/crates/gpui_macos/src/metal_renderer.rs @@ -7,9 +7,9 @@ use cocoa::{ quartzcore::AutoresizingMask, }; use gpui::{ - AtlasTextureId, BackdropBlur, Background, Bounds, ContentMask, DevicePixels, GlobalElementId, - MonochromeSprite, PaintSurface, Path, Point, PolychromeSprite, PrimitiveBatch, Quad, - RetainedLayer, RetainedLayerContentRevision, ScaledPixels, Scene, Shadow, Size, Surface, + AtlasTextureId, BackdropBlur, BackdropBlurPlan, Background, Bounds, ContentMask, DevicePixels, + GlobalElementId, MonochromeSprite, PaintSurface, Path, Point, PolychromeSprite, PrimitiveBatch, + Quad, RetainedLayer, RetainedLayerContentRevision, ScaledPixels, Scene, Shadow, Size, Surface, TransformationMatrix, Underline, point, size, }; #[cfg(any(test, feature = "test-support"))] @@ -46,9 +46,7 @@ const SHADERS_SOURCE_FILE: &str = include_str!(concat!(env!("OUT_DIR"), "/stitch // Use 4x MSAA, all devices support it. // https://developer.apple.com/documentation/metal/mtldevice/1433355-supportstexturesamplecount const PATH_SAMPLE_COUNT: u32 = 4; -const BACKDROP_BLUR_RADIUS_PER_LEVEL: f32 = 6.0; -const MAX_BACKDROP_BLUR_LEVELS: usize = 4; -const BACKDROP_BLUR_OFFSET: f32 = 1.0; +const MAX_BACKDROP_BLUR_LEVELS: usize = BackdropBlurPlan::MAX_PASSES; const BACKDROP_TEXTURE_SIZE_QUANTUM: i32 = 64; use crate::dispatch_semaphore::FrameSemaphore; @@ -208,7 +206,6 @@ pub(crate) struct MetalRenderer { backdrop_blur_level_sizes: Vec>, backdrop_blur_downsample_textures: Vec, backdrop_blur_upsample_textures: Vec, - backdrop_debug_last_scratch: Option<(i32, i32, u64, usize)>, retained_layers: HashMap, path_sample_count: u32, is_apple_gpu: bool, @@ -243,7 +240,7 @@ struct BackdropBlurParams { input_origin: [f32; 2], input_size: Size, texture_size: Size, - offset: f32, + sample_distance: f32, pad: f32, } @@ -389,7 +386,7 @@ impl MetalRenderer { "backdrop_blur_fragment", MTLPixelFormat::BGRA8Unorm, ); - let backdrop_blur_downsample_pipeline_state = build_pipeline_state( + let backdrop_blur_downsample_pipeline_state = build_no_blend_pipeline_state( &device, &library, "backdrop_blur_downsample", @@ -397,7 +394,7 @@ impl MetalRenderer { "backdrop_blur_downsample_fragment", MTLPixelFormat::BGRA8Unorm, ); - let backdrop_blur_upsample_pipeline_state = build_pipeline_state( + let backdrop_blur_upsample_pipeline_state = build_no_blend_pipeline_state( &device, &library, "backdrop_blur_upsample", @@ -492,7 +489,6 @@ impl MetalRenderer { backdrop_blur_level_sizes: Vec::new(), backdrop_blur_downsample_textures: Vec::new(), backdrop_blur_upsample_textures: Vec::new(), - backdrop_debug_last_scratch: None, retained_layers: HashMap::default(), path_sample_count: PATH_SAMPLE_COUNT, is_apple_gpu, @@ -578,7 +574,6 @@ impl MetalRenderer { self.backdrop_blur_level_sizes.clear(); self.backdrop_blur_downsample_textures.clear(); self.backdrop_blur_upsample_textures.clear(); - self.backdrop_debug_last_scratch = None; } fn ensure_path_intermediate_textures( @@ -646,7 +641,7 @@ impl MetalRenderer { ) -> Option> { let size = Self::quantize_backdrop_texture_size(size, max_size); if let Some(current_size) = self.backdrop_texture_size { - if current_size.width >= size.width && current_size.height >= size.height { + if Self::can_reuse_backdrop_texture(current_size, size, max_size) { return Some(current_size); } return self.create_backdrop_textures(Size { @@ -658,6 +653,17 @@ impl MetalRenderer { self.create_backdrop_textures(size) } + fn can_reuse_backdrop_texture( + current_size: Size, + required_size: Size, + max_size: Size, + ) -> bool { + current_size.width >= required_size.width + && current_size.height >= required_size.height + && current_size.width <= max_size.width + && current_size.height <= max_size.height + } + fn quantize_backdrop_texture_size( size: Size, max_size: Size, @@ -687,15 +693,6 @@ impl MetalRenderer { self.update_backdrop_blur_textures(size); self.backdrop_texture_size = Some(size); - if Self::backdrop_debug_enabled() { - eprintln!( - "[gpui backdrop blur] backdrop allocate {}x{} levels={} retained={:.1} MiB", - size.width.0, - size.height.0, - self.backdrop_blur_level_sizes.len(), - Self::bytes_to_mib(self.backdrop_texture_bytes()), - ); - } Some(size) } @@ -772,110 +769,6 @@ impl MetalRenderer { level_sizes } - fn backdrop_debug_enabled() -> bool { - std::env::var_os("GPUI_BACKDROP_BLUR_DEBUG").is_some() - } - - fn bytes_to_mib(bytes: u64) -> f64 { - bytes as f64 / 1024.0 / 1024.0 - } - - fn texture_bytes(size: Size) -> u64 { - size.width.0.max(0) as u64 * size.height.0.max(0) as u64 * 4 - } - - fn backdrop_texture_bytes_for_size(size: Size) -> u64 { - let levels = Self::backdrop_blur_level_sizes_for(size); - levels - .iter() - .skip(1) - .copied() - .map(Self::texture_bytes) - .sum::() - + levels.iter().copied().map(Self::texture_bytes).sum::() - } - - fn backdrop_texture_bytes(&self) -> u64 { - self.backdrop_texture_size - .map(Self::backdrop_texture_bytes_for_size) - .unwrap_or(0) - } - - fn backdrop_individual_area(blurs: &[BackdropBlur], viewport_size: Size) -> u64 { - let viewport_bounds = Bounds { - origin: point(ScaledPixels(0.0), ScaledPixels(0.0)), - size: size( - ScaledPixels::from(viewport_size.width), - ScaledPixels::from(viewport_size.height), - ), - }; - - blurs - .iter() - .map(|blur| { - let bounds = blur - .bounds - .dilate(Self::backdrop_blur_padding(blur.blur_radius.0)) - .intersect(&viewport_bounds); - if bounds.is_empty() { - return 0; - } - - let origin = bounds.origin.map(|component| component.floor()); - let bottom_right = bounds.bottom_right().map(|component| component.ceil()); - let bounds = Bounds::from_corners(origin, bottom_right); - DevicePixels::from(bounds.size.width).0.max(0) as u64 - * DevicePixels::from(bounds.size.height).0.max(0) as u64 - }) - .sum() - } - - fn log_backdrop_scratch( - &mut self, - blurs: &[BackdropBlur], - scratch_bounds: BackdropScratchBounds, - viewport_size: Size, - ) { - if !Self::backdrop_debug_enabled() { - return; - } - - let individual_area = Self::backdrop_individual_area(blurs, viewport_size); - let union_area = scratch_bounds.texture_size.width.0.max(0) as u64 - * scratch_bounds.texture_size.height.0.max(0) as u64; - let signature = ( - scratch_bounds.texture_size.width.0, - scratch_bounds.texture_size.height.0, - individual_area, - blurs.len(), - ); - if self.backdrop_debug_last_scratch == Some(signature) { - return; - } - self.backdrop_debug_last_scratch = Some(signature); - - let waste_ratio = if individual_area > 0 { - union_area as f64 / individual_area as f64 - } else { - 0.0 - }; - eprintln!( - "[gpui backdrop blur] batch blurs={} scratch={}x{} copy={:.1} MiB pyramid={:.1} MiB individual_area={} union_area={} waste={:.1}x viewport={}x{}", - blurs.len(), - scratch_bounds.texture_size.width.0, - scratch_bounds.texture_size.height.0, - Self::bytes_to_mib(Self::texture_bytes(scratch_bounds.texture_size)), - Self::bytes_to_mib(Self::backdrop_texture_bytes_for_size( - scratch_bounds.texture_size - )), - individual_area, - union_area, - waste_ratio, - viewport_size.width.0, - viewport_size.height.0, - ); - } - pub fn update_transparency(&self, transparent: bool) { self.layer.set_opaque(!transparent); } @@ -914,12 +807,6 @@ impl MetalRenderer { .acquire(&self.device, required_instance_buffer_size); if !has_backdrop_blurs && self.backdrop_texture_size.is_some() { - if Self::backdrop_debug_enabled() { - eprintln!( - "[gpui backdrop blur] no blur in scene; releasing backdrop textures={:.1} MiB", - Self::bytes_to_mib(self.backdrop_texture_bytes()), - ); - } self.discard_backdrop_textures(); } @@ -1542,21 +1429,6 @@ impl MetalRenderer { } }, ); - let frame_backdrop_scratch_bounds = Self::backdrop_scratch_bounds( - &scene.backdrop_blurs, - viewport_size, - ) - .and_then(|scratch_bounds| { - self.log_backdrop_scratch(&scene.backdrop_blurs, scratch_bounds, viewport_size); - self.ensure_backdrop_textures(scratch_bounds.texture_size, viewport_size) - .map(|texture_size| { - Self::fit_backdrop_scratch_bounds(scratch_bounds, texture_size, viewport_size) - }) - }); - let frame_backdrop_blur_level_sizes = frame_backdrop_scratch_bounds - .map(|scratch_bounds| Self::backdrop_blur_level_sizes_for(scratch_bounds.texture_size)) - .unwrap_or_default(); - for batch in scene.batches() { let ok = match batch { PrimitiveBatch::Shadows(range) => self.draw_shadows( @@ -1569,56 +1441,46 @@ impl MetalRenderer { PrimitiveBatch::BackdropBlurs(range) => { let blurs = &scene.backdrop_blurs[range]; command_encoder.end_encoding(); - if blurs.is_empty() { - command_encoder = new_command_encoder( - command_buffer, - target_texture, + let mut ok = true; + for blurs in Self::backdrop_blur_clusters(blurs, viewport_size) { + let Some(mut scratch_bounds) = + Self::backdrop_scratch_bounds(&blurs, viewport_size) + else { + continue; + }; + let Some(texture_size) = self.ensure_backdrop_textures( + scratch_bounds.texture_size, + Self::max_backdrop_texture_size(scratch_bounds, viewport_size), + ) else { + continue; + }; + scratch_bounds = Self::fit_backdrop_scratch_bounds( + scratch_bounds, + texture_size, viewport_size, - |color_attachment| { - color_attachment.set_load_action(metal::MTLLoadAction::Load); - }, ); - true - } else if let Some(scratch_bounds) = frame_backdrop_scratch_bounds { - let prepared_blurs = Self::prepare_backdrop_blurs(blurs, scratch_bounds); - let mut pass_groups = Vec::new(); - let mut start = 0; - while start < blurs.len() { - let passes = self.backdrop_blur_passes_for_radius( - blurs[start].blur_radius.0, - &frame_backdrop_blur_level_sizes, - ); - let mut end = start + 1; - while end < blurs.len() - && self.backdrop_blur_passes_for_radius( - blurs[end].blur_radius.0, - &frame_backdrop_blur_level_sizes, - ) == passes - { - end += 1; - } - pass_groups.push((start, end, passes)); - start = end; - } - - let needs_source_snapshot = pass_groups.len() > 1 - || pass_groups.iter().any(|(_, _, passes)| *passes == 0); - let source_snapshot = if needs_source_snapshot { - let Some(texture) = self.ensure_backdrop_source_texture() else { - anyhow::bail!("failed to create backdrop source texture"); + let level_sizes = + Self::backdrop_blur_level_sizes_for(scratch_bounds.texture_size); + let prepared_blurs = Self::prepare_backdrop_blurs(&blurs, scratch_bounds); + let plan_groups = Self::backdrop_blur_plan_groups(&blurs, &level_sizes); + + let source_snapshot = + if Self::backdrop_blur_needs_source_snapshot(&plan_groups) { + let Some(texture) = self.ensure_backdrop_source_texture() else { + anyhow::bail!("failed to create backdrop source texture"); + }; + self.render_texture_copy( + command_buffer, + target_texture, + &texture, + scratch_bounds.bounds.origin.x.0, + scratch_bounds.bounds.origin.y.0, + scratch_bounds.texture_size, + ); + Some(texture) + } else { + None }; - self.render_texture_copy( - command_buffer, - target_texture, - &texture, - scratch_bounds.bounds.origin.x.0, - scratch_bounds.bounds.origin.y.0, - scratch_bounds.texture_size, - ); - Some(texture) - } else { - None - }; let source_texture = source_snapshot .as_ref() @@ -1635,15 +1497,14 @@ impl MetalRenderer { point(ScaledPixels(0.0), ScaledPixels(0.0)); } - let mut ok = true; - for (start, end, passes) in pass_groups { - let blur_texture = self.render_backdrop_blur_texture_for_passes( + for (start, end, plan) in plan_groups { + let blur_texture = self.render_backdrop_blur_texture_for_plan( command_buffer, source_texture, source_texture_size, input_scratch_bounds, - passes, - &frame_backdrop_blur_level_sizes, + plan, + &level_sizes, source_snapshot.is_none(), ); command_encoder = new_command_encoder( @@ -1670,26 +1531,19 @@ impl MetalRenderer { break; } } - command_encoder = new_command_encoder( - command_buffer, - target_texture, - viewport_size, - |color_attachment| { - color_attachment.set_load_action(metal::MTLLoadAction::Load); - }, - ); - ok - } else { - command_encoder = new_command_encoder( - command_buffer, - target_texture, - viewport_size, - |color_attachment| { - color_attachment.set_load_action(metal::MTLLoadAction::Load); - }, - ); - true + if !ok { + break; + } } + command_encoder = new_command_encoder( + command_buffer, + target_texture, + viewport_size, + |color_attachment| { + color_attachment.set_load_action(metal::MTLLoadAction::Load); + }, + ); + ok } PrimitiveBatch::Quads(range) => self.draw_quads( &scene.quads[range], @@ -1917,6 +1771,86 @@ impl MetalRenderer { }) } + fn backdrop_source_bounds( + blur: &BackdropBlur, + viewport_size: Size, + ) -> Option> { + let viewport_bounds = Bounds { + origin: point(ScaledPixels(0.0), ScaledPixels(0.0)), + size: size( + ScaledPixels::from(viewport_size.width), + ScaledPixels::from(viewport_size.height), + ), + }; + let bounds = blur + .bounds + .dilate(Self::backdrop_blur_padding(blur.blur_radius.0)) + .intersect(&viewport_bounds); + if bounds.is_empty() { + return None; + } + + let origin = bounds.origin.map(|component| component.floor()); + let bottom_right = bounds.bottom_right().map(|component| component.ceil()); + Some(Bounds::from_corners(origin, bottom_right)) + } + + fn backdrop_blur_clusters( + blurs: &[BackdropBlur], + viewport_size: Size, + ) -> Vec> { + let mut clusters: Vec<(Bounds, Vec<(usize, BackdropBlur)>)> = Vec::new(); + for (blur_ix, blur) in blurs.iter().enumerate() { + let Some(mut cluster_bounds) = Self::backdrop_source_bounds(blur, viewport_size) else { + continue; + }; + let mut cluster_blurs = vec![(blur_ix, blur.clone())]; + + loop { + let overlapping_cluster_ixs = clusters + .iter() + .enumerate() + .filter_map(|(ix, (bounds, _))| { + bounds.intersects(&cluster_bounds).then_some(ix) + }) + .collect::>(); + if overlapping_cluster_ixs.is_empty() { + break; + } + + for ix in overlapping_cluster_ixs.into_iter().rev() { + let (bounds, mut blurs) = clusters.remove(ix); + cluster_bounds = cluster_bounds.union(&bounds); + cluster_blurs.append(&mut blurs); + } + } + + clusters.push((cluster_bounds, cluster_blurs)); + } + + clusters + .into_iter() + .map(|(_, mut blurs)| { + blurs.sort_by_key(|(ix, _)| *ix); + blurs.into_iter().map(|(_, blur)| blur).collect() + }) + .collect() + } + + fn max_backdrop_texture_size( + scratch_bounds: BackdropScratchBounds, + viewport_size: Size, + ) -> Size { + Size { + width: DevicePixels( + (viewport_size.width.0 - scratch_bounds.bounds.origin.x.0 as i32).max(0), + ), + height: DevicePixels( + (viewport_size.height.0 - scratch_bounds.bounds.origin.y.0 as i32).max(0), + ), + } + } + fn fit_backdrop_scratch_bounds( mut scratch_bounds: BackdropScratchBounds, texture_size: Size, @@ -1944,7 +1878,7 @@ impl MetalRenderer { } fn backdrop_blur_padding(radius: f32) -> ScaledPixels { - ScaledPixels((radius + BACKDROP_BLUR_RADIUS_PER_LEVEL * 2.0).ceil()) + ScaledPixels(BackdropBlurPlan::padding(radius)) } fn prepare_backdrop_blurs( @@ -1964,44 +1898,68 @@ impl MetalRenderer { .collect() } - fn backdrop_blur_passes_for_radius( - &self, + fn backdrop_blur_plan_for_radius( radius: f32, level_sizes: &[Size], - ) -> usize { - if radius <= 0.0 { - return 0; - } - let max_levels = level_sizes.len().saturating_sub(1); - if max_levels == 0 { - return 0; + ) -> BackdropBlurPlan { + BackdropBlurPlan::for_radius( + radius, + level_sizes + .len() + .saturating_sub(1) + .min(MAX_BACKDROP_BLUR_LEVELS), + ) + } + + fn backdrop_blur_plan_groups( + blurs: &[BackdropBlur], + level_sizes: &[Size], + ) -> Vec<(usize, usize, BackdropBlurPlan)> { + let mut groups = Vec::new(); + let mut start = 0; + while start < blurs.len() { + let plan = Self::backdrop_blur_plan_for_radius(blurs[start].blur_radius.0, level_sizes); + let mut end = start + 1; + while end < blurs.len() + && Self::backdrop_blur_plan_for_radius(blurs[end].blur_radius.0, level_sizes) + == plan + { + end += 1; + } + groups.push((start, end, plan)); + start = end; } - let passes = (radius / BACKDROP_BLUR_RADIUS_PER_LEVEL).ceil() as usize; - passes.clamp(1, max_levels) + groups } - fn render_backdrop_blur_texture_for_passes<'a>( + fn backdrop_blur_needs_source_snapshot( + plan_groups: &[(usize, usize, BackdropBlurPlan)], + ) -> bool { + plan_groups.len() > 1 || plan_groups.iter().any(|(_, _, plan)| plan.passes == 0) + } + + fn render_backdrop_blur_texture_for_plan<'a>( &'a self, command_buffer: &metal::CommandBufferRef, source_texture: &'a metal::TextureRef, source_texture_size: Size, scratch_bounds: BackdropScratchBounds, - passes: usize, + plan: BackdropBlurPlan, level_sizes: &[Size], source_texture_is_target: bool, ) -> Option<&'a metal::TextureRef> { - if passes == 0 { + if plan.passes == 0 { return (!source_texture_is_target).then_some(source_texture); } - if self.backdrop_blur_downsample_textures.len() < passes + if self.backdrop_blur_downsample_textures.len() < plan.passes || self.backdrop_blur_upsample_textures.is_empty() { return None; } let mut input_texture = source_texture; - for level in 0..passes { + for level in 0..plan.passes { let output_texture: &metal::TextureRef = &self.backdrop_blur_downsample_textures[level]; let input_size = level_sizes[level]; let output_size = level_sizes[level + 1]; @@ -2025,13 +1983,14 @@ impl MetalRenderer { input_size, texture_size, output_size, + plan.sample_distance, ); input_texture = output_texture; } let mut input_texture: &metal::TextureRef = - &self.backdrop_blur_downsample_textures[passes - 1]; - for level in (0..passes).rev() { + &self.backdrop_blur_downsample_textures[plan.passes - 1]; + for level in (0..plan.passes).rev() { let output_texture: &metal::TextureRef = &self.backdrop_blur_upsample_textures[level]; let input_size = level_sizes[level + 1]; let output_size = level_sizes[level]; @@ -2044,6 +2003,7 @@ impl MetalRenderer { input_size, input_size, output_size, + plan.sample_distance, ); input_texture = output_texture; } @@ -2063,6 +2023,7 @@ impl MetalRenderer { input_size: Size, texture_size: Size, output_size: Size, + sample_distance: f32, ) { let render_pass_descriptor = metal::RenderPassDescriptor::new(); let color_attachment = render_pass_descriptor @@ -2094,7 +2055,7 @@ impl MetalRenderer { input_origin, input_size, texture_size, - offset: BACKDROP_BLUR_OFFSET, + sample_distance, pad: 0.0, }; command_encoder.set_fragment_bytes( @@ -2941,6 +2902,36 @@ fn build_pipeline_state( .expect("could not create render pipeline state") } +fn build_no_blend_pipeline_state( + device: &metal::DeviceRef, + library: &metal::LibraryRef, + label: &str, + vertex_fn_name: &str, + fragment_fn_name: &str, + pixel_format: metal::MTLPixelFormat, +) -> metal::RenderPipelineState { + let vertex_fn = library + .get_function(vertex_fn_name, None) + .expect("error locating vertex function"); + let fragment_fn = library + .get_function(fragment_fn_name, None) + .expect("error locating fragment function"); + + let descriptor = metal::RenderPipelineDescriptor::new(); + descriptor.set_label(label); + descriptor.set_vertex_function(Some(vertex_fn.as_ref())); + descriptor.set_fragment_function(Some(fragment_fn.as_ref())); + descriptor + .color_attachments() + .object_at(0) + .unwrap() + .set_pixel_format(pixel_format); + + device + .new_render_pipeline_state(&descriptor) + .expect("could not create render pipeline state") +} + fn build_path_sprite_pipeline_state( device: &metal::DeviceRef, library: &metal::LibraryRef, @@ -3162,6 +3153,143 @@ mod tests { scene } + fn backdrop_blur(radius: f32) -> BackdropBlur { + backdrop_blur_with_bounds(0, 0.0, 100.0, radius) + } + + fn backdrop_blur_with_bounds(order: u32, x: f32, width: f32, radius: f32) -> BackdropBlur { + let bounds = Bounds::new( + Point::new(ScaledPixels(x), ScaledPixels(0.0)), + Size::new(ScaledPixels(width), ScaledPixels(10.0)), + ); + BackdropBlur { + order, + pad: 0, + bounds, + content_mask: ContentMask::new(bounds), + corner_radii: Corners::default(), + blur_radius: ScaledPixels(radius), + source_origin_x: 0.0, + source_origin_y: 0.0, + source_width: 1.0, + source_height: 1.0, + pad2: 0, + } + } + + #[test] + fn backdrop_blur_pyramid_supports_six_passes_and_small_textures_limit_plans() { + let full_level_sizes = MetalRenderer::backdrop_blur_level_sizes_for(Size::new( + DevicePixels(4096), + DevicePixels(4096), + )); + assert_eq!(full_level_sizes.len(), MAX_BACKDROP_BLUR_LEVELS + 1); + assert_eq!( + MetalRenderer::backdrop_blur_plan_for_radius(240.0, &full_level_sizes).passes, + MAX_BACKDROP_BLUR_LEVELS + ); + + let small_level_sizes = MetalRenderer::backdrop_blur_level_sizes_for(Size::new( + DevicePixels(8), + DevicePixels(8), + )); + assert_eq!(small_level_sizes.len(), 3); + assert_eq!( + MetalRenderer::backdrop_blur_plan_for_radius(240.0, &small_level_sizes).passes, + 2 + ); + } + + #[test] + fn backdrop_blur_groups_require_equal_sample_distance() { + let level_sizes = MetalRenderer::backdrop_blur_level_sizes_for(Size::new( + DevicePixels(256), + DevicePixels(256), + )); + let blurs = [backdrop_blur(8.0), backdrop_blur(8.0), backdrop_blur(9.0)]; + let groups = MetalRenderer::backdrop_blur_plan_groups(&blurs, &level_sizes); + + assert_eq!(groups.len(), 2); + assert_eq!((groups[0].0, groups[0].1, groups[0].2.passes), (0, 2, 1)); + assert_eq!((groups[1].0, groups[1].1, groups[1].2.passes), (2, 3, 1)); + assert_ne!(groups[0].2.sample_distance, groups[1].2.sample_distance); + } + + #[test] + fn backdrop_blur_clusters_use_transitive_dilated_overlap_and_preserve_order() { + let viewport_size = Size::new(DevicePixels(200), DevicePixels(100)); + let blurs = [ + backdrop_blur_with_bounds(7, 0.0, 10.0, 1.0), + backdrop_blur_with_bounds(3, 30.0, 10.0, 1.0), + backdrop_blur_with_bounds(9, 11.0, 18.0, 1.0), + backdrop_blur_with_bounds(11, 150.0, 10.0, 1.0), + backdrop_blur_with_bounds(13, 300.0, 10.0, 1.0), + ]; + + let clusters = MetalRenderer::backdrop_blur_clusters(&blurs, viewport_size); + + assert_eq!(clusters.len(), 2); + assert_eq!( + clusters[0] + .iter() + .map(|blur| blur.order) + .collect::>(), + vec![7, 3, 9] + ); + assert_eq!( + clusters[1] + .iter() + .map(|blur| blur.order) + .collect::>(), + vec![11] + ); + } + + #[test] + fn backdrop_blur_texture_reuse_respects_cluster_max_bounds() { + let required = Size::new(DevicePixels(64), DevicePixels(64)); + let max_size = Size::new(DevicePixels(128), DevicePixels(128)); + + assert!(MetalRenderer::can_reuse_backdrop_texture( + Size::new(DevicePixels(128), DevicePixels(128)), + required, + max_size, + )); + assert!(!MetalRenderer::can_reuse_backdrop_texture( + Size::new(DevicePixels(512), DevicePixels(512)), + required, + max_size, + )); + assert!(!MetalRenderer::can_reuse_backdrop_texture( + Size::new(DevicePixels(32), DevicePixels(128)), + required, + max_size, + )); + } + + #[test] + fn backdrop_blur_snapshot_predicate_preserves_fast_path() { + let positive_plan = BackdropBlurPlan { + passes: 1, + sample_distance: 1.0, + }; + + assert!(!MetalRenderer::backdrop_blur_needs_source_snapshot(&[( + 0, + 1, + positive_plan, + )])); + assert!(MetalRenderer::backdrop_blur_needs_source_snapshot(&[( + 0, + 1, + BackdropBlurPlan::IDENTITY, + )])); + assert!(MetalRenderer::backdrop_blur_needs_source_snapshot(&[ + (0, 1, positive_plan), + (1, 2, BackdropBlurPlan::IDENTITY), + ])); + } + #[test] fn retained_layer_with_backdrop_blur_is_excluded_from_metal_cache() { let bounds = Bounds::new( diff --git a/crates/gpui_macos/src/shaders.metal b/crates/gpui_macos/src/shaders.metal index 897a7b2..a32fa32 100644 --- a/crates/gpui_macos/src/shaders.metal +++ b/crates/gpui_macos/src/shaders.metal @@ -137,7 +137,7 @@ fragment float4 backdrop_blur_downsample_fragment( float2 texture_size = float2((float)params->texture_size.width, (float)params->texture_size.height); float2 texel = 1.0 / max(input_size, float2(1.0)); - float2 offset = texel * (params->offset + 0.5); + float2 offset = texel * 0.75; float3 rgb_sum = float3(0.0); float alpha_sum = 0.0; float4 sample = @@ -183,34 +183,60 @@ fragment float4 backdrop_blur_upsample_fragment( float2 texture_size = float2((float)params->texture_size.width, (float)params->texture_size.height); float2 texel = 1.0 / max(input_size, float2(1.0)); - float2 offset = texel * (params->offset + 0.5); + float2 offset = texel * params->sample_distance; float3 rgb_sum = float3(0.0); float alpha_sum = 0.0; + constexpr float axial_weight = 1.0 / 12.0; + constexpr float diagonal_weight = 1.0 / 6.0; float4 sample = + source_texture.sample(blur_sampler, backdrop_blur_sample_coord(input.uv + float2(-2.0 * offset.x, 0.0), input_origin, input_size, texture_size)); + if (sample.a > 0.0) { + rgb_sum += srgb_to_linear(sample.rgb / sample.a) * sample.a * axial_weight; + alpha_sum += sample.a * axial_weight; + } + sample = + source_texture.sample(blur_sampler, backdrop_blur_sample_coord(input.uv + float2(2.0 * offset.x, 0.0), input_origin, input_size, texture_size)); + if (sample.a > 0.0) { + rgb_sum += srgb_to_linear(sample.rgb / sample.a) * sample.a * axial_weight; + alpha_sum += sample.a * axial_weight; + } + sample = + source_texture.sample(blur_sampler, backdrop_blur_sample_coord(input.uv + float2(0.0, -2.0 * offset.y), input_origin, input_size, texture_size)); + if (sample.a > 0.0) { + rgb_sum += srgb_to_linear(sample.rgb / sample.a) * sample.a * axial_weight; + alpha_sum += sample.a * axial_weight; + } + sample = + source_texture.sample(blur_sampler, backdrop_blur_sample_coord(input.uv + float2(0.0, 2.0 * offset.y), input_origin, input_size, texture_size)); + if (sample.a > 0.0) { + rgb_sum += srgb_to_linear(sample.rgb / sample.a) * sample.a * axial_weight; + alpha_sum += sample.a * axial_weight; + } + sample = source_texture.sample(blur_sampler, backdrop_blur_sample_coord(input.uv + float2(-offset.x, -offset.y), input_origin, input_size, texture_size)); if (sample.a > 0.0) { - rgb_sum += srgb_to_linear(sample.rgb / sample.a) * sample.a; - alpha_sum += sample.a; + rgb_sum += srgb_to_linear(sample.rgb / sample.a) * sample.a * diagonal_weight; + alpha_sum += sample.a * diagonal_weight; } sample = source_texture.sample(blur_sampler, backdrop_blur_sample_coord(input.uv + float2(offset.x, -offset.y), input_origin, input_size, texture_size)); if (sample.a > 0.0) { - rgb_sum += srgb_to_linear(sample.rgb / sample.a) * sample.a; - alpha_sum += sample.a; + rgb_sum += srgb_to_linear(sample.rgb / sample.a) * sample.a * diagonal_weight; + alpha_sum += sample.a * diagonal_weight; } sample = source_texture.sample(blur_sampler, backdrop_blur_sample_coord(input.uv + float2(-offset.x, offset.y), input_origin, input_size, texture_size)); if (sample.a > 0.0) { - rgb_sum += srgb_to_linear(sample.rgb / sample.a) * sample.a; - alpha_sum += sample.a; + rgb_sum += srgb_to_linear(sample.rgb / sample.a) * sample.a * diagonal_weight; + alpha_sum += sample.a * diagonal_weight; } sample = source_texture.sample(blur_sampler, backdrop_blur_sample_coord(input.uv + float2(offset.x, offset.y), input_origin, input_size, texture_size)); if (sample.a > 0.0) { - rgb_sum += srgb_to_linear(sample.rgb / sample.a) * sample.a; - alpha_sum += sample.a; + rgb_sum += srgb_to_linear(sample.rgb / sample.a) * sample.a * diagonal_weight; + alpha_sum += sample.a * diagonal_weight; } - float alpha = alpha_sum * 0.25; + float alpha = alpha_sum; float safe_alpha = max(alpha_sum, 0.0001); float3 rgb = linear_to_srgb(rgb_sum / safe_alpha) * alpha; return float4(rgb, alpha); @@ -262,7 +288,9 @@ fragment float4 backdrop_blur_fragment( float alpha = clamp(0.5 - distance, 0.0, 1.0); alpha *= content_mask_alpha(input.position.xy, blur.content_mask.rounded_bounds, blur.content_mask.corner_radii); - return color * alpha; + float3 straight_rgb = + color.a > 0.0 ? color.rgb / color.a : float3(0.0); + return float4(straight_rgb, color.a * alpha); } vertex QuadVertexOutput quad_vertex(uint unit_vertex_id [[vertex_id]], diff --git a/crates/gpui_wgpu/src/shaders.wgsl b/crates/gpui_wgpu/src/shaders.wgsl index 0d51140..38895c2 100644 --- a/crates/gpui_wgpu/src/shaders.wgsl +++ b/crates/gpui_wgpu/src/shaders.wgsl @@ -1090,6 +1090,107 @@ fn fs_shadow(input: ShadowVarying) -> @location(0) vec4 { // --- backdrop blurs --- // +struct BackdropBlurPassParams { + input_size: vec2, + sample_distance: f32, + pad: f32, +} +@group(1) @binding(0) var b_backdrop_blur_pass_params: array; + +struct BackdropBlurPassVarying { + @builtin(position) position: vec4, + @location(0) uv: vec2, +} + +@vertex +fn vs_backdrop_blur_pass(@builtin(vertex_index) vertex_id: u32) -> BackdropBlurPassVarying { + let unit_vertex = vec2(f32(vertex_id & 1u), 0.5 * f32(vertex_id & 2u)); + var out = BackdropBlurPassVarying(); + out.position = vec4( + unit_vertex * vec2(2.0, -2.0) + vec2(-1.0, 1.0), + 0.0, + 1.0, + ); + out.uv = unit_vertex; + return out; +} + +fn backdrop_blur_downsample_color(uv: vec2) -> vec4 { + let params = b_backdrop_blur_pass_params[0]; + let texel = 1.0 / max(params.input_size, vec2(1.0)); + let offset = texel * 0.75; + let offsets = array, 4>( + vec2(-offset.x, -offset.y), + vec2(offset.x, -offset.y), + vec2(-offset.x, offset.y), + vec2(offset.x, offset.y), + ); + + var rgb_sum = vec3(0.0); + var alpha_sum = 0.0; + for (var i = 0u; i < 4u; i += 1u) { + let sample = textureSample(t_sprite, s_sprite, uv + offsets[i]); + if (sample.a > 0.0) { + rgb_sum += srgb_to_linear(sample.rgb / sample.a) * sample.a; + alpha_sum += sample.a; + } + } + + let alpha = alpha_sum * 0.25; + let rgb = linear_to_srgb(rgb_sum / max(alpha_sum, 0.0001)) * alpha; + return vec4(rgb, alpha); +} + +fn backdrop_blur_upsample_color(uv: vec2) -> vec4 { + let params = b_backdrop_blur_pass_params[0]; + let texel = 1.0 / max(params.input_size, vec2(1.0)); + let offset = texel * params.sample_distance; + let offsets = array, 8>( + vec2(-2.0 * offset.x, 0.0), + vec2(2.0 * offset.x, 0.0), + vec2(0.0, -2.0 * offset.y), + vec2(0.0, 2.0 * offset.y), + vec2(-offset.x, -offset.y), + vec2(offset.x, -offset.y), + vec2(-offset.x, offset.y), + vec2(offset.x, offset.y), + ); + let weights = array( + 1.0 / 12.0, + 1.0 / 12.0, + 1.0 / 12.0, + 1.0 / 12.0, + 1.0 / 6.0, + 1.0 / 6.0, + 1.0 / 6.0, + 1.0 / 6.0, + ); + + var rgb_sum = vec3(0.0); + var alpha_sum = 0.0; + for (var i = 0u; i < 8u; i += 1u) { + let sample = textureSample(t_sprite, s_sprite, uv + offsets[i]); + let weight = weights[i]; + if (sample.a > 0.0) { + rgb_sum += srgb_to_linear(sample.rgb / sample.a) * sample.a * weight; + alpha_sum += sample.a * weight; + } + } + + let rgb = linear_to_srgb(rgb_sum / max(alpha_sum, 0.0001)) * alpha_sum; + return vec4(rgb, alpha_sum); +} + +@fragment +fn fs_backdrop_blur_downsample(input: BackdropBlurPassVarying) -> @location(0) vec4 { + return backdrop_blur_downsample_color(input.uv); +} + +@fragment +fn fs_backdrop_blur_upsample(input: BackdropBlurPassVarying) -> @location(0) vec4 { + return backdrop_blur_upsample_color(input.uv); +} + struct BackdropBlur { order: u32, pad: u32, @@ -1142,45 +1243,16 @@ fn fs_backdrop_blur(input: BackdropBlurVarying) -> @location(0) vec4 { let source_origin = vec2(blur.source_origin_x, blur.source_origin_y); let source_size = max(vec2(blur.source_width, blur.source_height), vec2(1.0)); let uv = (input.position.xy - source_origin) / source_size; - let texel = 1.0 / source_size; - let radius_scale = max(blur.blur_radius / 3.0, 1.0); - let step = texel * radius_scale; - - var accum_linear = vec3(0.0); - var accum_alpha = 0.0; - var weight_sum = 0.0; - - let weights = array(0.204164, 0.123841, 0.123841, 0.123841, 0.123841, 0.0751136, 0.0751136, 0.0751136, 0.0751136); - let offsets = array, 9>( - vec2(0.0, 0.0), - vec2(step.x, 0.0), - vec2(-step.x, 0.0), - vec2(0.0, step.y), - vec2(0.0, -step.y), - vec2(step.x, step.y), - vec2(-step.x, step.y), - vec2(step.x, -step.y), - vec2(-step.x, -step.y), - ); - - for (var i = 0u; i < 9u; i += 1u) { - let w = weights[i]; - let sample = textureSample(t_sprite, s_sprite, uv + offsets[i]); - if (sample.a > 0.0) { - accum_linear += srgb_to_linear(sample.rgb / sample.a) * sample.a * w; - accum_alpha += sample.a * w; - } - weight_sum += w; + let color = textureSample(t_sprite, s_sprite, uv); + var straight_rgb = vec3(0.0); + if (color.a > 0.0) { + straight_rgb = color.rgb / color.a; } - let safe_alpha = max(accum_alpha, 0.0001); - let blurred_rgb = linear_to_srgb(accum_linear / safe_alpha); - let distance = quad_sdf(input.position.xy, blur.bounds, blur.corner_radii); let surface_alpha = saturate(0.5 - distance); - let alpha = (accum_alpha / max(weight_sum, 0.0001)) * surface_alpha; return apply_content_mask( - blend_color(vec4(blurred_rgb, 1.0), alpha), + blend_color(vec4(straight_rgb, color.a), surface_alpha), input.position.xy, blur.content_mask, ); diff --git a/crates/gpui_wgpu/src/wgpu_renderer.rs b/crates/gpui_wgpu/src/wgpu_renderer.rs index 3ecdba6..e3a35e7 100644 --- a/crates/gpui_wgpu/src/wgpu_renderer.rs +++ b/crates/gpui_wgpu/src/wgpu_renderer.rs @@ -1,10 +1,10 @@ use crate::{CompositorGpuHint, WgpuAtlas, WgpuContext}; use bytemuck::{Pod, Zeroable}; use gpui::{ - AtlasTextureId, BackdropBlur, Background, Bounds, ContentMask, DevicePixels, GlobalElementId, - GpuSpecs, MonochromeSprite, Path, Point, PolychromeSprite, PrimitiveBatch, Quad, RetainedLayer, - RetainedLayerContentRevision, ScaledPixels, Scene, Shadow, Size, SubpixelSprite, - TransformationMatrix, Underline, get_gamma_correction_ratios, + AtlasTextureId, BackdropBlur, BackdropBlurPlan, Background, Bounds, ContentMask, DevicePixels, + GlobalElementId, GpuSpecs, MonochromeSprite, Path, Point, PolychromeSprite, PrimitiveBatch, + Quad, RetainedLayer, RetainedLayerContentRevision, ScaledPixels, Scene, Shadow, Size, + SubpixelSprite, TransformationMatrix, Underline, get_gamma_correction_ratios, }; use log::warn; #[cfg(not(target_family = "wasm"))] @@ -17,7 +17,7 @@ use std::ops::Range; use std::rc::Rc; use std::sync::{Arc, Mutex}; -const BACKDROP_BLUR_RADIUS_PER_LEVEL: f32 = 6.0; +const MAX_BACKDROP_BLUR_LEVELS: usize = BackdropBlurPlan::MAX_PASSES; const BACKDROP_TEXTURE_SIZE_QUANTUM: i32 = 64; const INITIAL_INSTANCE_BUFFER_SIZE: u64 = 2 * 1024 * 1024; const INSTANCE_BUFFER_SIZE_BUCKET: u64 = 1024 * 1024; @@ -52,6 +52,10 @@ fn surface_usage(surface_usages: wgpu::TextureUsages) -> (wgpu::TextureUsages, b (usage, supports_copy_src) } +fn backdrop_blur_format(surface_format: wgpu::TextureFormat) -> wgpu::TextureFormat { + surface_format.remove_srgb_suffix() +} + #[derive(Default)] struct FrameFailureStreak { count: u32, @@ -173,6 +177,14 @@ struct BackdropScratchBounds { texture_size: Size, } +#[derive(Clone, Copy)] +#[repr(C)] +struct BackdropBlurPassParams { + input_size: [f32; 2], + sample_distance: f32, + pad: f32, +} + pub struct WgpuSurfaceConfig { pub size: Size, pub transparent: bool, @@ -208,6 +220,8 @@ struct WgpuPipelines { quads: wgpu::RenderPipeline, shadows: wgpu::RenderPipeline, backdrop_blurs: wgpu::RenderPipeline, + backdrop_blur_downsample: wgpu::RenderPipeline, + backdrop_blur_upsample: wgpu::RenderPipeline, path_rasterization: wgpu::RenderPipeline, paths: wgpu::RenderPipeline, underlines: wgpu::RenderPipeline, @@ -294,19 +308,46 @@ pub struct WgpuResources { backdrop_texture: Option, backdrop_view: Option, backdrop_size: Option>, + backdrop_blur_level_sizes: Vec>, + backdrop_blur_downsample_textures: Vec, + backdrop_blur_downsample_views: Vec, + backdrop_blur_upsample_textures: Vec, + backdrop_blur_upsample_views: Vec, retained_layers: HashMap, } impl WgpuResources { + fn discard_backdrop_resources(&mut self) { + self.backdrop_texture = None; + self.backdrop_view = None; + self.backdrop_size = None; + self.backdrop_blur_level_sizes.clear(); + self.backdrop_blur_downsample_textures.clear(); + self.backdrop_blur_downsample_views.clear(); + self.backdrop_blur_upsample_textures.clear(); + self.backdrop_blur_upsample_views.clear(); + } + + fn destroy_backdrop_resources(&mut self) { + if let Some(texture) = self.backdrop_texture.take() { + texture.destroy(); + } + for texture in self.backdrop_blur_downsample_textures.drain(..) { + texture.destroy(); + } + for texture in self.backdrop_blur_upsample_textures.drain(..) { + texture.destroy(); + } + self.discard_backdrop_resources(); + } + fn invalidate_cached_gpu_state(&mut self) { self.path_intermediate_texture = None; self.path_intermediate_view = None; self.path_msaa_texture = None; self.path_msaa_view = None; self.path_intermediate_size = None; - self.backdrop_texture = None; - self.backdrop_view = None; - self.backdrop_size = None; + self.discard_backdrop_resources(); self.retained_layers.clear(); } } @@ -691,6 +732,11 @@ impl WgpuRenderer { backdrop_texture: None, backdrop_view: None, backdrop_size: None, + backdrop_blur_level_sizes: Vec::new(), + backdrop_blur_downsample_textures: Vec::new(), + backdrop_blur_downsample_views: Vec::new(), + backdrop_blur_upsample_textures: Vec::new(), + backdrop_blur_upsample_views: Vec::new(), retained_layers: HashMap::default(), }; @@ -956,6 +1002,31 @@ impl WgpuRenderer { &[Some(color_target.clone())], 1, ); + let backdrop_blur_target = wgpu::ColorTargetState { + format: backdrop_blur_format(surface_format), + blend: None, + write_mask: wgpu::ColorWrites::ALL, + }; + let backdrop_blur_downsample = create_pipeline( + "backdrop_blur_downsample", + "vs_backdrop_blur_pass", + "fs_backdrop_blur_downsample", + &layouts.globals, + &layouts.instances_with_texture, + wgpu::PrimitiveTopology::TriangleStrip, + &[Some(backdrop_blur_target.clone())], + 1, + ); + let backdrop_blur_upsample = create_pipeline( + "backdrop_blur_upsample", + "vs_backdrop_blur_pass", + "fs_backdrop_blur_upsample", + &layouts.globals, + &layouts.instances_with_texture, + wgpu::PrimitiveTopology::TriangleStrip, + &[Some(backdrop_blur_target)], + 1, + ); let path_rasterization = create_pipeline( "path_rasterization", @@ -1091,6 +1162,8 @@ impl WgpuRenderer { quads, shadows, backdrop_blurs, + backdrop_blur_downsample, + backdrop_blur_upsample, path_rasterization, paths, underlines, @@ -1150,6 +1223,30 @@ impl WgpuRenderer { (texture, view) } + fn create_backdrop_blur_texture( + device: &wgpu::Device, + format: wgpu::TextureFormat, + size: Size, + label: &'static str, + ) -> (wgpu::Texture, wgpu::TextureView) { + let texture = device.create_texture(&wgpu::TextureDescriptor { + label: Some(label), + size: wgpu::Extent3d { + width: size.width.0.max(1) as u32, + height: size.height.0.max(1) as u32, + depth_or_array_layers: 1, + }, + mip_level_count: 1, + sample_count: 1, + dimension: wgpu::TextureDimension::D2, + format, + usage: wgpu::TextureUsages::RENDER_ATTACHMENT | wgpu::TextureUsages::TEXTURE_BINDING, + view_formats: &[], + }); + let view = texture.create_view(&wgpu::TextureViewDescriptor::default()); + (texture, view) + } + fn create_msaa_if_needed( device: &wgpu::Device, format: wgpu::TextureFormat, @@ -1227,12 +1324,7 @@ impl WgpuRenderer { self.path_msaa_view = None; self.path_intermediate_size = None; - if let Some(ref texture) = self.backdrop_texture { - texture.destroy(); - } - self.backdrop_texture = None; - self.backdrop_view = None; - self.backdrop_size = None; + self.destroy_backdrop_resources(); } } @@ -1309,11 +1401,7 @@ impl WgpuRenderer { if let Some(current_size) = self.backdrop_size && self.backdrop_texture.is_some() { - if current_size.width >= size.width - && current_size.height >= size.height - && current_size.width <= max_size.width - && current_size.height <= max_size.height - { + if Self::can_reuse_backdrop_texture(current_size, size, max_size) { return Some(current_size); } return self.create_backdrop_resources(Size { @@ -1325,6 +1413,17 @@ impl WgpuRenderer { self.create_backdrop_resources(size) } + fn can_reuse_backdrop_texture( + current_size: Size, + required_size: Size, + max_size: Size, + ) -> bool { + current_size.width >= required_size.width + && current_size.height >= required_size.height + && current_size.width <= max_size.width + && current_size.height <= max_size.height + } + fn quantize_backdrop_texture_size( size: Size, max_size: Size, @@ -1351,9 +1450,7 @@ impl WgpuRenderer { size: Size, ) -> Option> { if size.width.0 <= 0 || size.height.0 <= 0 { - self.backdrop_texture = None; - self.backdrop_view = None; - self.backdrop_size = None; + self.discard_backdrop_resources(); return None; } @@ -1363,9 +1460,40 @@ impl WgpuRenderer { size.width.0 as u32, size.height.0 as u32, ); + let level_sizes = Self::backdrop_blur_level_sizes_for(size); + let blur_format = backdrop_blur_format(self.surface_config.format); + let mut downsample_textures = Vec::with_capacity(level_sizes.len().saturating_sub(1)); + let mut downsample_views = Vec::with_capacity(level_sizes.len().saturating_sub(1)); + for &level_size in level_sizes.iter().skip(1) { + let (texture, view) = Self::create_backdrop_blur_texture( + &self.device, + blur_format, + level_size, + "backdrop_blur_downsample", + ); + downsample_textures.push(texture); + downsample_views.push(view); + } + let mut upsample_textures = Vec::with_capacity(level_sizes.len()); + let mut upsample_views = Vec::with_capacity(level_sizes.len()); + for &level_size in &level_sizes { + let (texture, view) = Self::create_backdrop_blur_texture( + &self.device, + blur_format, + level_size, + "backdrop_blur_upsample", + ); + upsample_textures.push(texture); + upsample_views.push(view); + } self.backdrop_texture = Some(backdrop_texture); self.backdrop_view = Some(backdrop_view); self.backdrop_size = Some(size); + self.backdrop_blur_level_sizes = level_sizes; + self.backdrop_blur_downsample_textures = downsample_textures; + self.backdrop_blur_downsample_views = downsample_views; + self.backdrop_blur_upsample_textures = upsample_textures; + self.backdrop_blur_upsample_views = upsample_views; Some(size) } @@ -1459,9 +1587,7 @@ impl WgpuRenderer { .iter() .any(|blur| Self::backdrop_source_bounds(blur, viewport_size).is_some()) { - self.backdrop_texture = None; - self.backdrop_view = None; - self.backdrop_size = None; + self.discard_backdrop_resources(); } let frame = match self.surface.get_current_texture() { @@ -2035,6 +2161,10 @@ impl WgpuRenderer { }; let prepared_blurs = Self::prepare_backdrop_blurs(&blurs, scratch_bounds); + let plan_groups = Self::backdrop_blur_plan_groups( + &blurs, + &self.backdrop_blur_level_sizes, + ); drop(pass); encoder.copy_texture_to_texture( wgpu::TexelCopyTextureInfo { @@ -2059,8 +2189,54 @@ impl WgpuRenderer { depth_or_array_layers: 1, }, ); + + for (start, end, plan) in plan_groups { + if !self.render_backdrop_blur_for_plan( + encoder, + plan, + instance_offset, + ) { + ok = false; + break; + } + let blur_view = if plan.passes == 0 { + self.backdrop_view.as_ref() + } else { + self.backdrop_blur_upsample_views.first() + }; + let Some(blur_view) = blur_view else { + ok = false; + break; + }; + let mut blur_pass = + encoder.begin_render_pass(&wgpu::RenderPassDescriptor { + label: Some("main_pass_backdrop"), + color_attachments: &[Some( + wgpu::RenderPassColorAttachment { + view: target_view, + resolve_target: None, + ops: wgpu::Operations { + load: wgpu::LoadOp::Load, + store: wgpu::StoreOp::Store, + }, + depth_slice: None, + }, + )], + depth_stencil_attachment: None, + ..Default::default() + }); + if !self.draw_backdrop_blurs( + &prepared_blurs[start..end], + blur_view, + instance_offset, + &mut blur_pass, + ) { + ok = false; + break; + } + } pass = encoder.begin_render_pass(&wgpu::RenderPassDescriptor { - label: Some("main_pass_backdrop"), + label: Some("main_pass_after_backdrop"), color_attachments: &[Some(wgpu::RenderPassColorAttachment { view: target_view, resolve_target: None, @@ -2073,11 +2249,6 @@ impl WgpuRenderer { depth_stencil_attachment: None, ..Default::default() }); - ok = self.draw_backdrop_blurs( - &prepared_blurs, - instance_offset, - &mut pass, - ); if !ok { break; } @@ -2289,12 +2460,10 @@ impl WgpuRenderer { fn draw_backdrop_blurs( &self, blurs: &[BackdropBlur], + backdrop_view: &wgpu::TextureView, instance_offset: &mut u64, pass: &mut wgpu::RenderPass<'_>, ) -> bool { - let Some(backdrop_view) = self.backdrop_view.as_ref() else { - return false; - }; let data = unsafe { Self::instance_bytes(blurs) }; self.draw_instances_with_texture( data, @@ -2306,6 +2475,106 @@ impl WgpuRenderer { ) } + fn render_backdrop_blur_for_plan( + &self, + encoder: &mut wgpu::CommandEncoder, + plan: BackdropBlurPlan, + instance_offset: &mut u64, + ) -> bool { + if plan.passes == 0 { + return self.backdrop_view.is_some(); + } + if self.backdrop_blur_downsample_views.len() < plan.passes + || self.backdrop_blur_upsample_views.is_empty() + { + return false; + } + + for level in 0..plan.passes { + let input_view = if level == 0 { + let Some(view) = self.backdrop_view.as_ref() else { + return false; + }; + view + } else { + &self.backdrop_blur_downsample_views[level - 1] + }; + let params = BackdropBlurPassParams { + input_size: [ + self.backdrop_blur_level_sizes[level].width.0 as f32, + self.backdrop_blur_level_sizes[level].height.0 as f32, + ], + sample_distance: plan.sample_distance, + pad: 0.0, + }; + if !self.draw_backdrop_blur_pass( + encoder, + ¶ms, + input_view, + &self.backdrop_blur_downsample_views[level], + &self.pipelines.backdrop_blur_downsample, + instance_offset, + ) { + return false; + } + } + + for level in (0..plan.passes).rev() { + let input_view = if level == plan.passes - 1 { + &self.backdrop_blur_downsample_views[level] + } else { + &self.backdrop_blur_upsample_views[level + 1] + }; + let params = BackdropBlurPassParams { + input_size: [ + self.backdrop_blur_level_sizes[level + 1].width.0 as f32, + self.backdrop_blur_level_sizes[level + 1].height.0 as f32, + ], + sample_distance: plan.sample_distance, + pad: 0.0, + }; + if !self.draw_backdrop_blur_pass( + encoder, + ¶ms, + input_view, + &self.backdrop_blur_upsample_views[level], + &self.pipelines.backdrop_blur_upsample, + instance_offset, + ) { + return false; + } + } + + true + } + + fn draw_backdrop_blur_pass( + &self, + encoder: &mut wgpu::CommandEncoder, + params: &BackdropBlurPassParams, + input_view: &wgpu::TextureView, + output_view: &wgpu::TextureView, + pipeline: &wgpu::RenderPipeline, + instance_offset: &mut u64, + ) -> bool { + let mut pass = encoder.begin_render_pass(&wgpu::RenderPassDescriptor { + label: Some("backdrop_blur_pass"), + color_attachments: &[Some(wgpu::RenderPassColorAttachment { + view: output_view, + resolve_target: None, + ops: wgpu::Operations { + load: wgpu::LoadOp::Clear(wgpu::Color::TRANSPARENT), + store: wgpu::StoreOp::Store, + }, + depth_slice: None, + })], + depth_stencil_attachment: None, + ..Default::default() + }); + let data = unsafe { Self::instance_bytes(std::slice::from_ref(params)) }; + self.draw_instances_with_texture(data, 1, input_view, pipeline, instance_offset, &mut pass) + } + fn draw_underlines( &self, underlines: &[Underline], @@ -2625,7 +2894,64 @@ impl WgpuRenderer { } fn backdrop_blur_padding(radius: f32) -> ScaledPixels { - ScaledPixels((radius + BACKDROP_BLUR_RADIUS_PER_LEVEL * 2.0).ceil()) + ScaledPixels(BackdropBlurPlan::padding(radius)) + } + + fn backdrop_blur_level_sizes_for(size: Size) -> Vec> { + let mut level_sizes = Vec::new(); + if size.width.0 <= 0 || size.height.0 <= 0 { + return level_sizes; + } + + level_sizes.push(size); + let mut level_size = size; + for _ in 0..MAX_BACKDROP_BLUR_LEVELS { + let next_width = level_size.width.0 / 2; + let next_height = level_size.height.0 / 2; + if next_width < 2 || next_height < 2 { + break; + } + level_size = Size { + width: DevicePixels(next_width), + height: DevicePixels(next_height), + }; + level_sizes.push(level_size); + } + level_sizes + } + + fn backdrop_blur_plan_for_radius( + radius: f32, + level_sizes: &[Size], + ) -> BackdropBlurPlan { + BackdropBlurPlan::for_radius( + radius, + level_sizes + .len() + .saturating_sub(1) + .min(MAX_BACKDROP_BLUR_LEVELS), + ) + } + + fn backdrop_blur_plan_groups( + blurs: &[BackdropBlur], + level_sizes: &[Size], + ) -> Vec<(usize, usize, BackdropBlurPlan)> { + let mut groups = Vec::new(); + let mut start = 0; + while start < blurs.len() { + let plan = Self::backdrop_blur_plan_for_radius(blurs[start].blur_radius.0, level_sizes); + let mut end = start + 1; + while end < blurs.len() + && Self::backdrop_blur_plan_for_radius(blurs[end].blur_radius.0, level_sizes) + == plan + { + end += 1; + } + groups.push((start, end, plan)); + start = end; + } + groups } fn prepare_backdrop_blurs( @@ -2787,6 +3113,9 @@ impl WgpuRenderer { } PrimitiveBatch::BackdropBlurs(range) => { self.add_instance_bytes::(&mut size, range.len()); + for _ in 0..range.len() * MAX_BACKDROP_BLUR_LEVELS * 2 { + self.add_instance_bytes::(&mut size, 1); + } } PrimitiveBatch::Paths(range) => { for path in &scene.paths[range] { @@ -2993,6 +3322,172 @@ mod tests { (scene, layer) } + fn backdrop_blur(radius: f32) -> BackdropBlur { + backdrop_blur_with_bounds(0, 0.0, 100.0, radius) + } + + fn backdrop_blur_with_bounds(order: u32, x: f32, width: f32, radius: f32) -> BackdropBlur { + let bounds = Bounds::new( + Point::new(ScaledPixels(x), ScaledPixels(0.0)), + Size::new(ScaledPixels(width), ScaledPixels(10.0)), + ); + BackdropBlur { + order, + pad: 0, + bounds, + content_mask: ContentMask::new(bounds), + corner_radii: Corners::default(), + blur_radius: ScaledPixels(radius), + source_origin_x: 0.0, + source_origin_y: 0.0, + source_width: 1.0, + source_height: 1.0, + pad2: 0, + } + } + + #[test] + fn backdrop_blur_pyramid_supports_six_passes_and_small_textures_limit_plans() { + let full_level_sizes = WgpuRenderer::backdrop_blur_level_sizes_for(Size::new( + DevicePixels(4096), + DevicePixels(4096), + )); + assert_eq!(full_level_sizes.len(), MAX_BACKDROP_BLUR_LEVELS + 1); + assert_eq!( + WgpuRenderer::backdrop_blur_plan_for_radius(240.0, &full_level_sizes).passes, + MAX_BACKDROP_BLUR_LEVELS + ); + + let small_level_sizes = WgpuRenderer::backdrop_blur_level_sizes_for(Size::new( + DevicePixels(8), + DevicePixels(8), + )); + assert_eq!(small_level_sizes.len(), 3); + assert_eq!( + WgpuRenderer::backdrop_blur_plan_for_radius(240.0, &small_level_sizes).passes, + 2 + ); + } + + #[test] + fn backdrop_blur_groups_require_equal_sample_distance() { + let level_sizes = WgpuRenderer::backdrop_blur_level_sizes_for(Size::new( + DevicePixels(256), + DevicePixels(256), + )); + let blurs = [backdrop_blur(8.0), backdrop_blur(8.0), backdrop_blur(9.0)]; + let groups = WgpuRenderer::backdrop_blur_plan_groups(&blurs, &level_sizes); + + assert_eq!(groups.len(), 2); + assert_eq!((groups[0].0, groups[0].1, groups[0].2.passes), (0, 2, 1)); + assert_eq!((groups[1].0, groups[1].1, groups[1].2.passes), (2, 3, 1)); + assert_ne!(groups[0].2.sample_distance, groups[1].2.sample_distance); + } + + #[test] + fn backdrop_blur_clusters_use_transitive_dilated_overlap_and_preserve_order() { + let viewport_size = Size::new(DevicePixels(200), DevicePixels(100)); + let blurs = [ + backdrop_blur_with_bounds(7, 0.0, 10.0, 1.0), + backdrop_blur_with_bounds(3, 30.0, 10.0, 1.0), + backdrop_blur_with_bounds(9, 11.0, 18.0, 1.0), + backdrop_blur_with_bounds(11, 150.0, 10.0, 1.0), + backdrop_blur_with_bounds(13, 300.0, 10.0, 1.0), + ]; + + let clusters = WgpuRenderer::backdrop_blur_clusters(&blurs, viewport_size); + + assert_eq!(clusters.len(), 2); + assert_eq!( + clusters[0] + .iter() + .map(|blur| blur.order) + .collect::>(), + vec![7, 3, 9] + ); + assert_eq!( + clusters[1] + .iter() + .map(|blur| blur.order) + .collect::>(), + vec![11] + ); + } + + #[test] + fn backdrop_blur_pass_params_match_wgsl_storage_layout() { + assert_eq!(std::mem::size_of::(), 16); + assert_eq!(std::mem::align_of::(), 4); + } + + #[test] + fn backdrop_blur_intermediates_use_linear_formats() { + assert_eq!( + backdrop_blur_format(wgpu::TextureFormat::Bgra8UnormSrgb), + wgpu::TextureFormat::Bgra8Unorm + ); + assert_eq!( + backdrop_blur_format(wgpu::TextureFormat::Rgba8UnormSrgb), + wgpu::TextureFormat::Rgba8Unorm + ); + assert_eq!( + backdrop_blur_format(wgpu::TextureFormat::Bgra8Unorm), + wgpu::TextureFormat::Bgra8Unorm + ); + assert_eq!( + backdrop_blur_format(wgpu::TextureFormat::Rgba8Unorm), + wgpu::TextureFormat::Rgba8Unorm + ); + assert_eq!( + backdrop_blur_format(wgpu::TextureFormat::Rgba16Float), + wgpu::TextureFormat::Rgba16Float + ); + } + + #[test] + fn backdrop_blur_texture_reuse_respects_cluster_max_bounds() { + let required = Size::new(DevicePixels(64), DevicePixels(64)); + let max_size = Size::new(DevicePixels(128), DevicePixels(128)); + + assert!(WgpuRenderer::can_reuse_backdrop_texture( + Size::new(DevicePixels(128), DevicePixels(128)), + required, + max_size, + )); + assert!(!WgpuRenderer::can_reuse_backdrop_texture( + Size::new(DevicePixels(512), DevicePixels(512)), + required, + max_size, + )); + assert!(!WgpuRenderer::can_reuse_backdrop_texture( + Size::new(DevicePixels(32), DevicePixels(128)), + required, + max_size, + )); + } + + #[test] + fn shaders_parse_and_validate() { + let module = + wgpu::naga::front::wgsl::parse_str(include_str!("shaders.wgsl")).expect("parse WGSL"); + wgpu::naga::valid::Validator::new( + wgpu::naga::valid::ValidationFlags::all(), + wgpu::naga::valid::Capabilities::all(), + ) + .validate(&module) + .expect("validate WGSL"); + } + + #[test] + fn backdrop_blur_shader_uses_fixed_downsample_and_weighted_tent_upsample() { + let source = include_str!("shaders.wgsl"); + assert!(source.contains("let offset = texel * 0.75;")); + assert!(source.contains("let offsets = array, 8>(")); + assert!(source.contains("let weights = array(")); + assert!(source.contains("1.0 / 12.0,")); + assert!(source.contains("1.0 / 6.0,")); + } + #[test] fn select_present_mode_accepts_auto_modes_without_surface_advertising_them() { assert_eq!( diff --git a/crates/gpui_windows/src/directx_renderer.rs b/crates/gpui_windows/src/directx_renderer.rs index 1c529c8..2790054 100644 --- a/crates/gpui_windows/src/directx_renderer.rs +++ b/crates/gpui_windows/src/directx_renderer.rs @@ -40,9 +40,7 @@ pub(crate) const DISABLE_DIRECT_COMPOSITION: &str = "GPUI_DISABLE_DIRECT_COMPOSI const RENDER_TARGET_FORMAT: DXGI_FORMAT = DXGI_FORMAT_B8G8R8A8_UNORM; // This configuration is used for MSAA rendering on paths only, and it's guaranteed to be supported by DirectX 11. const PATH_MULTISAMPLE_COUNT: u32 = 4; -const BACKDROP_BLUR_RADIUS_PER_LEVEL: f32 = 6.0; -const MAX_BACKDROP_BLUR_LEVELS: usize = 4; -const BACKDROP_BLUR_OFFSET: f32 = 1.0; +const MAX_BACKDROP_BLUR_LEVELS: usize = BackdropBlurPlan::MAX_PASSES; const BACKDROP_TEXTURE_SIZE_QUANTUM: i32 = 64; const MAX_STRUCTURED_BUFFER_BYTES: usize = 256 * 1024 * 1024; @@ -599,24 +597,28 @@ impl DirectXRenderer { let prepared_blurs = Self::prepare_backdrop_blurs(&blurs, scratch_bounds); self.copy_render_target_to_backdrop(scratch_bounds)?; - let mut current_passes = None; + let max_backdrop_blur_levels = self.max_backdrop_blur_levels(); + let mut current_plan = None; let mut current_blur_srv = None; let mut start = 0; while start < blurs.len() { - let passes = - self.backdrop_blur_passes_for_radius(blurs[start].blur_radius.0); + let plan = Self::backdrop_blur_plan_for_radius( + blurs[start].blur_radius.0, + max_backdrop_blur_levels, + ); let mut end = start + 1; while end < blurs.len() - && self - .backdrop_blur_passes_for_radius(blurs[end].blur_radius.0) - == passes + && Self::backdrop_blur_plan_for_radius( + blurs[end].blur_radius.0, + max_backdrop_blur_levels, + ) == plan { end += 1; } - if current_passes != Some(passes) { + if current_plan != Some(plan) { current_blur_srv = - self.run_backdrop_blur_passes_for_passes(passes)?; - current_passes = Some(passes); + self.run_backdrop_blur_passes(plan)?; + current_plan = Some(plan); } self.draw_backdrop_blurs( &prepared_blurs[start..end], @@ -1133,7 +1135,7 @@ impl DirectXRenderer { } fn backdrop_blur_padding(radius: f32) -> ScaledPixels { - ScaledPixels((radius + BACKDROP_BLUR_RADIUS_PER_LEVEL * 2.0).ceil()) + ScaledPixels(BackdropBlurPlan::padding(radius)) } fn prepare_backdrop_blurs( @@ -1153,24 +1155,17 @@ impl DirectXRenderer { .collect() } - fn backdrop_blur_passes_for_radius(&self, radius: f32) -> usize { - if radius <= 0.0 { - return 0; - } - let resources = match self.resources.as_ref() { - Some(resources) => resources, - None => return 0, - }; - let max_levels = resources - .backdrop_blur + fn max_backdrop_blur_levels(&self) -> usize { + self.resources .as_ref() + .and_then(|resources| resources.backdrop_blur.as_ref()) .map(|blur| blur.level_sizes.len().saturating_sub(1)) - .unwrap_or(0); - if max_levels == 0 { - return 0; - } - let passes = (radius / BACKDROP_BLUR_RADIUS_PER_LEVEL).ceil() as usize; - passes.clamp(1, max_levels) + .unwrap_or(0) + .min(MAX_BACKDROP_BLUR_LEVELS) + } + + fn backdrop_blur_plan_for_radius(radius: f32, max_levels: usize) -> BackdropBlurPlan { + BackdropBlurPlan::for_radius(radius, max_levels.min(MAX_BACKDROP_BLUR_LEVELS)) } fn draw_backdrop_blurs( @@ -1202,9 +1197,9 @@ impl DirectXRenderer { ) } - fn run_backdrop_blur_passes_for_passes( + fn run_backdrop_blur_passes( &mut self, - passes: usize, + plan: BackdropBlurPlan, ) -> Result> { let devices = self.devices.as_ref().context("devices missing")?; let resources = self.resources.as_ref().context("resources missing")?; @@ -1212,15 +1207,17 @@ impl DirectXRenderer { .backdrop_blur .as_ref() .context("backdrop blur resources missing")?; - if passes == 0 { + if plan.passes == 0 { return Ok(resources.backdrop_srv.clone()); } - if backdrop_blur.downsample_srvs.len() < passes || backdrop_blur.upsample_srvs.is_empty() { + if backdrop_blur.downsample_srvs.len() < plan.passes + || backdrop_blur.upsample_srvs.is_empty() + { return Ok(resources.backdrop_srv.clone()); } let mut input_srv = resources.backdrop_srv.clone(); - for level in 0..passes { + for level in 0..plan.passes { let input_size = backdrop_blur.level_sizes[level]; let output_size = backdrop_blur.level_sizes[level + 1]; let output_view = backdrop_blur.downsample_views[level].as_ref(); @@ -1234,7 +1231,7 @@ impl DirectXRenderer { }; let params = BackdropBlurParams { input_size: [input_size.0 as f32, input_size.1 as f32], - offset: BACKDROP_BLUR_OFFSET, + sample_distance: plan.sample_distance, pad: 0.0, }; Self::draw_backdrop_blur_pass( @@ -1250,8 +1247,8 @@ impl DirectXRenderer { input_srv = backdrop_blur.downsample_srvs[level].clone(); } - let mut input_srv = backdrop_blur.downsample_srvs[passes - 1].clone(); - for level in (0..passes).rev() { + let mut input_srv = backdrop_blur.downsample_srvs[plan.passes - 1].clone(); + for level in (0..plan.passes).rev() { let input_size = backdrop_blur.level_sizes[level + 1]; let output_size = backdrop_blur.level_sizes[level]; let output_view = backdrop_blur.upsample_views[level].as_ref(); @@ -1265,7 +1262,7 @@ impl DirectXRenderer { }; let params = BackdropBlurParams { input_size: [input_size.0 as f32, input_size.1 as f32], - offset: BACKDROP_BLUR_OFFSET, + sample_distance: plan.sample_distance, pad: 0.0, }; Self::draw_backdrop_blur_pass( @@ -1897,14 +1894,14 @@ impl DirectXRenderPipelines { "backdrop_blur_downsample_pipeline", ShaderModule::BackdropBlurDownsample, 4, - create_blend_state(device)?, + create_blend_state_without_blending(device)?, )?; let backdrop_blur_upsample_pipeline = PipelineState::new( device, "backdrop_blur_upsample_pipeline", ShaderModule::BackdropBlurUpsample, 4, - create_blend_state(device)?, + create_blend_state_without_blending(device)?, )?; let quad_pipeline = PipelineState::new( device, @@ -2252,7 +2249,7 @@ struct GlobalParams { #[repr(C)] struct BackdropBlurParams { input_size: [f32; 2], - offset: f32, + sample_distance: f32, pad: f32, } @@ -2808,6 +2805,17 @@ fn create_blend_state(device: &ID3D11Device) -> Result { } } +#[inline] +fn create_blend_state_without_blending(device: &ID3D11Device) -> Result { + let mut desc = D3D11_BLEND_DESC::default(); + desc.RenderTarget[0].RenderTargetWriteMask = D3D11_COLOR_WRITE_ENABLE_ALL.0 as u8; + unsafe { + let mut state = None; + device.CreateBlendState(&desc, Some(&mut state))?; + Ok(state.unwrap()) + } +} + #[inline] fn create_blend_state_for_subpixel_rendering(device: &ID3D11Device) -> Result { let mut desc = D3D11_BLEND_DESC::default(); @@ -3379,17 +3387,85 @@ mod amd { #[cfg(test)] mod tests { - use gpui::{Bounds, ContentMask, Corners, ScaledPixels, point, size}; + use gpui::{ + BackdropBlur, Bounds, ContentMask, Corners, Point, ScaledPixels, Size, point, size, + }; use super::{ - GlobalParams, RetainedLayerClip, retained_layer_clip, rounded_backdrop_rebuild_requested, + BackdropBlurParams, DirectXRenderer, GlobalParams, RetainedLayerClip, + backdrop_blur_level_sizes, retained_layer_clip, rounded_backdrop_rebuild_requested, }; + fn backdrop_blur_with_bounds(order: u32, x: f32, width: f32, radius: f32) -> BackdropBlur { + let bounds = Bounds::new( + Point::new(ScaledPixels(x), ScaledPixels(0.0)), + Size::new(ScaledPixels(width), ScaledPixels(10.0)), + ); + BackdropBlur { + order, + pad: 0, + bounds, + content_mask: ContentMask::new(bounds), + corner_radii: Corners::default(), + blur_radius: ScaledPixels(radius), + source_origin_x: 0.0, + source_origin_y: 0.0, + source_width: 1.0, + source_height: 1.0, + pad2: 0, + } + } + #[test] fn global_params_preserve_hlsl_constant_buffer_alignment() { assert_eq!(std::mem::size_of::(), 48); } + #[test] + fn backdrop_blur_params_preserve_hlsl_buffer_layout() { + assert_eq!(std::mem::size_of::(), 16); + } + + #[test] + fn backdrop_blur_pyramid_caps_levels_and_groups_by_full_plan() { + assert_eq!(backdrop_blur_level_sizes(4_096, 4_096).len(), 7); + assert_eq!(backdrop_blur_level_sizes(3, 3), vec![(3, 3)]); + + let first = DirectXRenderer::backdrop_blur_plan_for_radius(8.0, 6); + let second = DirectXRenderer::backdrop_blur_plan_for_radius(9.0, 6); + assert_eq!(first.passes, second.passes); + assert_ne!(first, second); + } + + #[test] + fn backdrop_blur_clusters_use_transitive_dilated_overlap_and_preserve_order() { + let blurs = [ + backdrop_blur_with_bounds(7, 0.0, 10.0, 1.0), + backdrop_blur_with_bounds(3, 30.0, 10.0, 1.0), + backdrop_blur_with_bounds(9, 11.0, 18.0, 1.0), + backdrop_blur_with_bounds(11, 150.0, 10.0, 1.0), + backdrop_blur_with_bounds(13, 300.0, 10.0, 1.0), + ]; + + let clusters = DirectXRenderer::backdrop_blur_clusters(&blurs, 200, 100); + + assert_eq!(clusters.len(), 2); + assert_eq!( + clusters[0] + .iter() + .map(|blur| blur.order) + .collect::>(), + vec![7, 3, 9] + ); + assert_eq!( + clusters[1] + .iter() + .map(|blur| blur.order) + .collect::>(), + vec![11] + ); + } + #[test] fn rounded_backdrop_rebuild_uses_retained_configuration() { assert!(!rounded_backdrop_rebuild_requested(None)); diff --git a/crates/gpui_windows/src/shaders.hlsl b/crates/gpui_windows/src/shaders.hlsl index 0117c5a..a2399cd 100644 --- a/crates/gpui_windows/src/shaders.hlsl +++ b/crates/gpui_windows/src/shaders.hlsl @@ -128,12 +128,12 @@ float4 distance_from_clip_rect_transformed(float2 unit_vertex, Bounds bounds, Bo // Convert linear RGB to sRGB float3 linear_to_srgb(float3 color) { - return pow(color, float3(2.2, 2.2, 2.2)); + return pow(color, float3(1.0 / 2.2, 1.0 / 2.2, 1.0 / 2.2)); } // Convert sRGB to linear RGB float3 srgb_to_linear(float3 color) { - return pow(color, float3(1.0 / 2.2, 1.0 / 2.2, 1.0 / 2.2)); + return pow(color, float3(2.2, 2.2, 2.2)); } /// Hsla to linear RGBA conversion. @@ -564,7 +564,7 @@ struct BackdropBlur { struct BackdropBlurParams { float2 input_size; - float offset; + float sample_distance; float pad; }; @@ -608,7 +608,7 @@ BackdropBlurPassVertexOutput backdrop_blur_upsample_vertex(uint vertex_id: SV_Ve float4 backdrop_blur_downsample_fragment(BackdropBlurPassVertexOutput input) : SV_Target { BackdropBlurParams params = backdrop_blur_params[0]; float2 texel = 1.0 / max(params.input_size, float2(1.0, 1.0)); - float2 offset = texel * (params.offset + 0.5); + float2 offset = texel * 0.75; float3 rgb_sum = float3(0.0, 0.0, 0.0); float alpha_sum = 0.0; float4 sample = t_sprite.Sample(s_sprite, input.uv + float2(-offset.x, -offset.y)); @@ -640,30 +640,51 @@ float4 backdrop_blur_downsample_fragment(BackdropBlurPassVertexOutput input) : S float4 backdrop_blur_upsample_fragment(BackdropBlurPassVertexOutput input) : SV_Target { BackdropBlurParams params = backdrop_blur_params[0]; float2 texel = 1.0 / max(params.input_size, float2(1.0, 1.0)); - float2 offset = texel * (params.offset + 0.5); + float2 diagonal_offset = texel * params.sample_distance; + float2 axial_offset = diagonal_offset * 2.0; float3 rgb_sum = float3(0.0, 0.0, 0.0); float alpha_sum = 0.0; - float4 sample = t_sprite.Sample(s_sprite, input.uv + float2(-offset.x, -offset.y)); + float4 sample = t_sprite.Sample(s_sprite, input.uv + float2(-axial_offset.x, 0.0)); if (sample.a > 0.0) { - rgb_sum += srgb_to_linear(sample.rgb / sample.a) * sample.a; - alpha_sum += sample.a; + rgb_sum += srgb_to_linear(sample.rgb / sample.a) * sample.a / 12.0; + alpha_sum += sample.a / 12.0; } - sample = t_sprite.Sample(s_sprite, input.uv + float2(offset.x, -offset.y)); + sample = t_sprite.Sample(s_sprite, input.uv + float2(axial_offset.x, 0.0)); if (sample.a > 0.0) { - rgb_sum += srgb_to_linear(sample.rgb / sample.a) * sample.a; - alpha_sum += sample.a; + rgb_sum += srgb_to_linear(sample.rgb / sample.a) * sample.a / 12.0; + alpha_sum += sample.a / 12.0; } - sample = t_sprite.Sample(s_sprite, input.uv + float2(-offset.x, offset.y)); + sample = t_sprite.Sample(s_sprite, input.uv + float2(0.0, -axial_offset.y)); if (sample.a > 0.0) { - rgb_sum += srgb_to_linear(sample.rgb / sample.a) * sample.a; - alpha_sum += sample.a; + rgb_sum += srgb_to_linear(sample.rgb / sample.a) * sample.a / 12.0; + alpha_sum += sample.a / 12.0; } - sample = t_sprite.Sample(s_sprite, input.uv + float2(offset.x, offset.y)); + sample = t_sprite.Sample(s_sprite, input.uv + float2(0.0, axial_offset.y)); if (sample.a > 0.0) { - rgb_sum += srgb_to_linear(sample.rgb / sample.a) * sample.a; - alpha_sum += sample.a; + rgb_sum += srgb_to_linear(sample.rgb / sample.a) * sample.a / 12.0; + alpha_sum += sample.a / 12.0; } - float alpha = alpha_sum * 0.25; + sample = t_sprite.Sample(s_sprite, input.uv + float2(-diagonal_offset.x, -diagonal_offset.y)); + if (sample.a > 0.0) { + rgb_sum += srgb_to_linear(sample.rgb / sample.a) * sample.a / 6.0; + alpha_sum += sample.a / 6.0; + } + sample = t_sprite.Sample(s_sprite, input.uv + float2(diagonal_offset.x, -diagonal_offset.y)); + if (sample.a > 0.0) { + rgb_sum += srgb_to_linear(sample.rgb / sample.a) * sample.a / 6.0; + alpha_sum += sample.a / 6.0; + } + sample = t_sprite.Sample(s_sprite, input.uv + float2(-diagonal_offset.x, diagonal_offset.y)); + if (sample.a > 0.0) { + rgb_sum += srgb_to_linear(sample.rgb / sample.a) * sample.a / 6.0; + alpha_sum += sample.a / 6.0; + } + sample = t_sprite.Sample(s_sprite, input.uv + float2(diagonal_offset.x, diagonal_offset.y)); + if (sample.a > 0.0) { + rgb_sum += srgb_to_linear(sample.rgb / sample.a) * sample.a / 6.0; + alpha_sum += sample.a / 6.0; + } + float alpha = alpha_sum; float safe_alpha = max(alpha_sum, 0.0001); float3 rgb = linear_to_srgb(rgb_sum / safe_alpha) * alpha; return float4(rgb, alpha); @@ -698,7 +719,8 @@ float4 backdrop_blur_fragment(BackdropBlurFragmentInput input) : SV_Target { float distance = quad_sdf(input.position.xy, blur.bounds, blur.corner_radii); float alpha = saturate(0.5 - distance); alpha *= content_mask_alpha(input.position.xy, blur.content_mask); - return color * alpha; + float3 rgb = color.a > 0.0 ? color.rgb / color.a : float3(0.0, 0.0, 0.0); + return float4(rgb, color.a * alpha); } QuadVertexOutput quad_vertex(uint vertex_id: SV_VertexID, uint quad_id: SV_InstanceID) { From bfbe11afed55781fe519b1beba1387cd38b949df Mon Sep 17 00:00:00 2001 From: Aditya Sharma Date: Mon, 27 Jul 2026 21:43:05 -0700 Subject: [PATCH 2/4] fix(rendering): address backdrop blur review findings Centralize pure blur planning, preserve source epochs, prevent duplicate sRGB decoding, and retain safe texture allocations across backend clusters. --- crates/gpui/src/scene.rs | 248 ++++++++++++++- crates/gpui_macos/src/metal_renderer.rs | 272 +--------------- crates/gpui_wgpu/src/wgpu_renderer.rs | 334 ++++---------------- crates/gpui_windows/src/directx_renderer.rs | 241 ++++---------- 4 files changed, 373 insertions(+), 722 deletions(-) diff --git a/crates/gpui/src/scene.rs b/crates/gpui/src/scene.rs index 92cf8bd..3c411e4 100644 --- a/crates/gpui/src/scene.rs +++ b/crates/gpui/src/scene.rs @@ -5,8 +5,9 @@ use schemars::JsonSchema; use serde::{Deserialize, Serialize}; use crate::{ - AtlasTextureId, AtlasTile, Background, Bounds, ContentMask, Corners, Edges, GlobalElementId, - Hsla, Pixels, Point, Radians, ScaledPixels, Size, bounds_tree::BoundsTree, point, + AtlasTextureId, AtlasTile, Background, Bounds, ContentMask, Corners, DevicePixels, Edges, + GlobalElementId, Hsla, Pixels, Point, Radians, ScaledPixels, Size, bounds_tree::BoundsTree, + point, }; use std::{ fmt::Debug, @@ -829,6 +830,135 @@ impl PartialEq for BackdropBlurPlan { impl Eq for BackdropBlurPlan {} +/// Returns the padded source extent for a backdrop blur radius. +#[doc(hidden)] +pub fn backdrop_blur_padding(radius: f32) -> ScaledPixels { + ScaledPixels(BackdropBlurPlan::padding(radius)) +} + +/// Returns the texture dimensions for each available backdrop blur pyramid level. +#[doc(hidden)] +pub fn backdrop_blur_level_sizes_for(size: Size) -> Vec> { + let mut level_sizes = Vec::new(); + if size.width.0 <= 0 || size.height.0 <= 0 { + return level_sizes; + } + + level_sizes.push(size); + let mut level_size = size; + for _ in 0..BackdropBlurPlan::MAX_PASSES { + let next_width = level_size.width.0 / 2; + let next_height = level_size.height.0 / 2; + if next_width < 2 || next_height < 2 { + break; + } + level_size = Size { + width: DevicePixels(next_width), + height: DevicePixels(next_height), + }; + level_sizes.push(level_size); + } + level_sizes +} + +/// Builds a backdrop blur plan limited to the available pyramid passes. +#[doc(hidden)] +pub fn backdrop_blur_plan_for_radius(radius: f32, available_passes: usize) -> BackdropBlurPlan { + BackdropBlurPlan::for_radius(radius, available_passes) +} + +/// Groups adjacent backdrop blurs that share one complete blur plan. +#[doc(hidden)] +pub fn backdrop_blur_plan_groups( + blurs: &[BackdropBlur], + available_passes: usize, +) -> Vec<(usize, usize, BackdropBlurPlan)> { + let mut groups = Vec::new(); + let mut start = 0; + while start < blurs.len() { + let plan = backdrop_blur_plan_for_radius(blurs[start].blur_radius.0, available_passes); + let mut end = start + 1; + while end < blurs.len() + && backdrop_blur_plan_for_radius(blurs[end].blur_radius.0, available_passes) == plan + { + end += 1; + } + groups.push((start, end, plan)); + start = end; + } + groups +} + +/// Returns whether an allocated backdrop texture can fit the required region. +#[doc(hidden)] +pub fn can_reuse_backdrop_texture( + current_size: Size, + required_size: Size, +) -> bool { + current_size.width >= required_size.width && current_size.height >= required_size.height +} + +/// Returns the viewport-clipped source bounds required for one backdrop blur. +#[doc(hidden)] +pub fn backdrop_source_bounds( + blur: &BackdropBlur, + viewport_size: Size, +) -> Option> { + let viewport_bounds = Bounds { + origin: point(ScaledPixels(0.0), ScaledPixels(0.0)), + size: Size { + width: ScaledPixels::from(viewport_size.width), + height: ScaledPixels::from(viewport_size.height), + }, + }; + let bounds = blur + .bounds + .dilate(backdrop_blur_padding(blur.blur_radius.0)) + .intersect(&viewport_bounds); + if bounds.is_empty() { + return None; + } + + let origin = bounds.origin.map(|component| component.floor()); + let bottom_right = bounds.bottom_right().map(|component| component.ceil()); + Some(Bounds::from_corners(origin, bottom_right)) +} + +/// Clusters backdrop blurs whose padded source regions overlap. +/// +/// Intervening clusters join a merged cluster so each source snapshot covers a +/// contiguous draw-order epoch. +#[doc(hidden)] +pub fn backdrop_blur_clusters( + blurs: &[BackdropBlur], + viewport_size: Size, +) -> Vec> { + let mut clusters: Vec<(Bounds, Vec)> = Vec::new(); + for blur in blurs { + let Some(mut cluster_bounds) = backdrop_source_bounds(blur, viewport_size) else { + continue; + }; + let mut cluster_blurs = vec![blur.clone()]; + + while let Some(first_overlapping_cluster_ix) = clusters + .iter() + .position(|(bounds, _)| bounds.intersects(&cluster_bounds)) + { + let mut preceding_blurs = Vec::new(); + for (bounds, mut blurs) in clusters.drain(first_overlapping_cluster_ix..) { + cluster_bounds = cluster_bounds.union(&bounds); + preceding_blurs.append(&mut blurs); + } + preceding_blurs.append(&mut cluster_blurs); + cluster_blurs = preceding_blurs; + } + + clusters.push((cluster_bounds, cluster_blurs)); + } + + clusters.into_iter().map(|(_, blurs)| blurs).collect() +} + #[derive(Debug, Clone)] #[repr(C)] #[expect(missing_docs)] @@ -1269,14 +1399,26 @@ mod tests { } fn test_backdrop_blur(order: DrawOrder) -> BackdropBlur { - let bounds = test_bounds(); + test_backdrop_blur_with_bounds(order, 0.0, 10.0, 12.0) + } + + fn test_backdrop_blur_with_bounds( + order: DrawOrder, + x: f32, + width: f32, + radius: f32, + ) -> BackdropBlur { + let bounds = Bounds::new( + point(ScaledPixels(x), ScaledPixels(0.0)), + size(ScaledPixels(width), ScaledPixels(10.0)), + ); BackdropBlur { order, pad: 0, bounds, content_mask: ContentMask::new(bounds), corner_radii: Corners::all(ScaledPixels(2.0)), - blur_radius: ScaledPixels(12.0), + blur_radius: ScaledPixels(radius), source_origin_x: 0.0, source_origin_y: 0.0, source_width: 1.0, @@ -1383,6 +1525,104 @@ mod tests { assert_eq!(BackdropBlurPlan::padding(f32::NAN), 0.0); } + #[test] + fn backdrop_blur_pyramid_limits_plans_to_available_levels() { + let full_level_sizes = + backdrop_blur_level_sizes_for(size(DevicePixels(4096), DevicePixels(4096))); + assert_eq!(full_level_sizes.len(), BackdropBlurPlan::MAX_PASSES + 1); + assert_eq!( + backdrop_blur_plan_for_radius(240.0, full_level_sizes.len() - 1).passes, + BackdropBlurPlan::MAX_PASSES + ); + + let small_level_sizes = + backdrop_blur_level_sizes_for(size(DevicePixels(8), DevicePixels(8))); + assert_eq!(small_level_sizes.len(), 3); + assert_eq!( + backdrop_blur_plan_for_radius(240.0, small_level_sizes.len() - 1).passes, + 2 + ); + } + + #[test] + fn backdrop_blur_groups_require_equal_sample_distance() { + let blurs = [ + test_backdrop_blur_with_bounds(0, 0.0, 10.0, 8.0), + test_backdrop_blur_with_bounds(1, 0.0, 10.0, 8.0), + test_backdrop_blur_with_bounds(2, 0.0, 10.0, 9.0), + ]; + let groups = backdrop_blur_plan_groups(&blurs, BackdropBlurPlan::MAX_PASSES); + + assert_eq!(groups.len(), 2); + assert_eq!((groups[0].0, groups[0].1, groups[0].2.passes), (0, 2, 1)); + assert_eq!((groups[1].0, groups[1].1, groups[1].2.passes), (2, 3, 1)); + assert_ne!(groups[0].2.sample_distance, groups[1].2.sample_distance); + } + + #[test] + fn backdrop_blur_texture_reuse_depends_only_on_required_size() { + let required = size(DevicePixels(64), DevicePixels(64)); + + assert!(can_reuse_backdrop_texture( + size(DevicePixels(512), DevicePixels(512)), + required, + )); + assert!(!can_reuse_backdrop_texture( + size(DevicePixels(32), DevicePixels(128)), + required, + )); + } + + #[test] + fn backdrop_blur_clusters_preserve_interleaved_source_order() { + let blurs = [ + test_backdrop_blur_with_bounds(7, 0.0, 10.0, 1.0), + test_backdrop_blur_with_bounds(3, 50.0, 10.0, 1.0), + test_backdrop_blur_with_bounds(9, 15.0, 10.0, 1.0), + ]; + + let clusters = backdrop_blur_clusters(&blurs, size(DevicePixels(200), DevicePixels(100))); + + assert_eq!(clusters.len(), 1); + assert_eq!( + clusters + .iter() + .flatten() + .map(|blur| blur.order) + .collect::>(), + vec![7, 3, 9] + ); + } + + #[test] + fn backdrop_blur_clusters_merge_adjacent_transitive_overlaps() { + let blurs = [ + test_backdrop_blur_with_bounds(7, 0.0, 10.0, 1.0), + test_backdrop_blur_with_bounds(3, 30.0, 10.0, 1.0), + test_backdrop_blur_with_bounds(9, 11.0, 18.0, 1.0), + test_backdrop_blur_with_bounds(11, 150.0, 10.0, 1.0), + test_backdrop_blur_with_bounds(13, 300.0, 10.0, 1.0), + ]; + + let clusters = backdrop_blur_clusters(&blurs, size(DevicePixels(200), DevicePixels(100))); + + assert_eq!(clusters.len(), 2); + assert_eq!( + clusters[0] + .iter() + .map(|blur| blur.order) + .collect::>(), + vec![7, 3, 9] + ); + assert_eq!( + clusters[1] + .iter() + .map(|blur| blur.order) + .collect::>(), + vec![11] + ); + } + #[test] fn padded_bool32_has_initialized_u32_representation() { assert_eq!(size_of::(), size_of::()); diff --git a/crates/gpui_macos/src/metal_renderer.rs b/crates/gpui_macos/src/metal_renderer.rs index d04818e..ef6fdb3 100644 --- a/crates/gpui_macos/src/metal_renderer.rs +++ b/crates/gpui_macos/src/metal_renderer.rs @@ -10,7 +10,8 @@ use gpui::{ AtlasTextureId, BackdropBlur, BackdropBlurPlan, Background, Bounds, ContentMask, DevicePixels, GlobalElementId, MonochromeSprite, PaintSurface, Path, Point, PolychromeSprite, PrimitiveBatch, Quad, RetainedLayer, RetainedLayerContentRevision, ScaledPixels, Scene, Shadow, Size, Surface, - TransformationMatrix, Underline, point, size, + TransformationMatrix, Underline, backdrop_blur_clusters, backdrop_blur_level_sizes_for, + backdrop_blur_padding, backdrop_blur_plan_groups, can_reuse_backdrop_texture, point, size, }; #[cfg(any(test, feature = "test-support"))] use image::RgbaImage; @@ -46,7 +47,6 @@ const SHADERS_SOURCE_FILE: &str = include_str!(concat!(env!("OUT_DIR"), "/stitch // Use 4x MSAA, all devices support it. // https://developer.apple.com/documentation/metal/mtldevice/1433355-supportstexturesamplecount const PATH_SAMPLE_COUNT: u32 = 4; -const MAX_BACKDROP_BLUR_LEVELS: usize = BackdropBlurPlan::MAX_PASSES; const BACKDROP_TEXTURE_SIZE_QUANTUM: i32 = 64; use crate::dispatch_semaphore::FrameSemaphore; @@ -641,7 +641,7 @@ impl MetalRenderer { ) -> Option> { let size = Self::quantize_backdrop_texture_size(size, max_size); if let Some(current_size) = self.backdrop_texture_size { - if Self::can_reuse_backdrop_texture(current_size, size, max_size) { + if can_reuse_backdrop_texture(current_size, size) { return Some(current_size); } return self.create_backdrop_textures(Size { @@ -653,17 +653,6 @@ impl MetalRenderer { self.create_backdrop_textures(size) } - fn can_reuse_backdrop_texture( - current_size: Size, - required_size: Size, - max_size: Size, - ) -> bool { - current_size.width >= required_size.width - && current_size.height >= required_size.height - && current_size.width <= max_size.width - && current_size.height <= max_size.height - } - fn quantize_backdrop_texture_size( size: Size, max_size: Size, @@ -720,7 +709,7 @@ impl MetalRenderer { self.backdrop_blur_downsample_textures.clear(); self.backdrop_blur_upsample_textures.clear(); - self.backdrop_blur_level_sizes = Self::backdrop_blur_level_sizes_for(size); + self.backdrop_blur_level_sizes = backdrop_blur_level_sizes_for(size); if self.backdrop_blur_level_sizes.is_empty() { return; } @@ -746,29 +735,6 @@ impl MetalRenderer { } } - fn backdrop_blur_level_sizes_for(size: Size) -> Vec> { - let mut level_sizes = Vec::new(); - if size.width.0 <= 0 || size.height.0 <= 0 { - return level_sizes; - } - - level_sizes.push(size); - let mut level_size = size; - for _ in 0..MAX_BACKDROP_BLUR_LEVELS { - let next_width = level_size.width.0 / 2; - let next_height = level_size.height.0 / 2; - if next_width < 2 || next_height < 2 { - break; - } - level_size = Size { - width: DevicePixels(next_width), - height: DevicePixels(next_height), - }; - level_sizes.push(level_size); - } - level_sizes - } - pub fn update_transparency(&self, transparent: bool) { self.layer.set_opaque(!transparent); } @@ -1442,7 +1408,7 @@ impl MetalRenderer { let blurs = &scene.backdrop_blurs[range]; command_encoder.end_encoding(); let mut ok = true; - for blurs in Self::backdrop_blur_clusters(blurs, viewport_size) { + for blurs in backdrop_blur_clusters(blurs, viewport_size) { let Some(mut scratch_bounds) = Self::backdrop_scratch_bounds(&blurs, viewport_size) else { @@ -1459,10 +1425,10 @@ impl MetalRenderer { texture_size, viewport_size, ); - let level_sizes = - Self::backdrop_blur_level_sizes_for(scratch_bounds.texture_size); + let level_sizes = self.backdrop_blur_level_sizes.clone(); let prepared_blurs = Self::prepare_backdrop_blurs(&blurs, scratch_bounds); - let plan_groups = Self::backdrop_blur_plan_groups(&blurs, &level_sizes); + let plan_groups = + backdrop_blur_plan_groups(&blurs, level_sizes.len().saturating_sub(1)); let source_snapshot = if Self::backdrop_blur_needs_source_snapshot(&plan_groups) { @@ -1738,12 +1704,12 @@ impl MetalRenderer { let mut bounds = blurs .first()? .bounds - .dilate(Self::backdrop_blur_padding(blurs.first()?.blur_radius.0)); + .dilate(backdrop_blur_padding(blurs.first()?.blur_radius.0)); for blur in blurs.iter().skip(1) { bounds = bounds.union( &blur .bounds - .dilate(Self::backdrop_blur_padding(blur.blur_radius.0)), + .dilate(backdrop_blur_padding(blur.blur_radius.0)), ); } @@ -1771,72 +1737,6 @@ impl MetalRenderer { }) } - fn backdrop_source_bounds( - blur: &BackdropBlur, - viewport_size: Size, - ) -> Option> { - let viewport_bounds = Bounds { - origin: point(ScaledPixels(0.0), ScaledPixels(0.0)), - size: size( - ScaledPixels::from(viewport_size.width), - ScaledPixels::from(viewport_size.height), - ), - }; - let bounds = blur - .bounds - .dilate(Self::backdrop_blur_padding(blur.blur_radius.0)) - .intersect(&viewport_bounds); - if bounds.is_empty() { - return None; - } - - let origin = bounds.origin.map(|component| component.floor()); - let bottom_right = bounds.bottom_right().map(|component| component.ceil()); - Some(Bounds::from_corners(origin, bottom_right)) - } - - fn backdrop_blur_clusters( - blurs: &[BackdropBlur], - viewport_size: Size, - ) -> Vec> { - let mut clusters: Vec<(Bounds, Vec<(usize, BackdropBlur)>)> = Vec::new(); - for (blur_ix, blur) in blurs.iter().enumerate() { - let Some(mut cluster_bounds) = Self::backdrop_source_bounds(blur, viewport_size) else { - continue; - }; - let mut cluster_blurs = vec![(blur_ix, blur.clone())]; - - loop { - let overlapping_cluster_ixs = clusters - .iter() - .enumerate() - .filter_map(|(ix, (bounds, _))| { - bounds.intersects(&cluster_bounds).then_some(ix) - }) - .collect::>(); - if overlapping_cluster_ixs.is_empty() { - break; - } - - for ix in overlapping_cluster_ixs.into_iter().rev() { - let (bounds, mut blurs) = clusters.remove(ix); - cluster_bounds = cluster_bounds.union(&bounds); - cluster_blurs.append(&mut blurs); - } - } - - clusters.push((cluster_bounds, cluster_blurs)); - } - - clusters - .into_iter() - .map(|(_, mut blurs)| { - blurs.sort_by_key(|(ix, _)| *ix); - blurs.into_iter().map(|(_, blur)| blur).collect() - }) - .collect() - } - fn max_backdrop_texture_size( scratch_bounds: BackdropScratchBounds, viewport_size: Size, @@ -1877,10 +1777,6 @@ impl MetalRenderer { scratch_bounds } - fn backdrop_blur_padding(radius: f32) -> ScaledPixels { - ScaledPixels(BackdropBlurPlan::padding(radius)) - } - fn prepare_backdrop_blurs( blurs: &[BackdropBlur], scratch_bounds: BackdropScratchBounds, @@ -1898,40 +1794,6 @@ impl MetalRenderer { .collect() } - fn backdrop_blur_plan_for_radius( - radius: f32, - level_sizes: &[Size], - ) -> BackdropBlurPlan { - BackdropBlurPlan::for_radius( - radius, - level_sizes - .len() - .saturating_sub(1) - .min(MAX_BACKDROP_BLUR_LEVELS), - ) - } - - fn backdrop_blur_plan_groups( - blurs: &[BackdropBlur], - level_sizes: &[Size], - ) -> Vec<(usize, usize, BackdropBlurPlan)> { - let mut groups = Vec::new(); - let mut start = 0; - while start < blurs.len() { - let plan = Self::backdrop_blur_plan_for_radius(blurs[start].blur_radius.0, level_sizes); - let mut end = start + 1; - while end < blurs.len() - && Self::backdrop_blur_plan_for_radius(blurs[end].blur_radius.0, level_sizes) - == plan - { - end += 1; - } - groups.push((start, end, plan)); - start = end; - } - groups - } - fn backdrop_blur_needs_source_snapshot( plan_groups: &[(usize, usize, BackdropBlurPlan)], ) -> bool { @@ -3153,120 +3015,6 @@ mod tests { scene } - fn backdrop_blur(radius: f32) -> BackdropBlur { - backdrop_blur_with_bounds(0, 0.0, 100.0, radius) - } - - fn backdrop_blur_with_bounds(order: u32, x: f32, width: f32, radius: f32) -> BackdropBlur { - let bounds = Bounds::new( - Point::new(ScaledPixels(x), ScaledPixels(0.0)), - Size::new(ScaledPixels(width), ScaledPixels(10.0)), - ); - BackdropBlur { - order, - pad: 0, - bounds, - content_mask: ContentMask::new(bounds), - corner_radii: Corners::default(), - blur_radius: ScaledPixels(radius), - source_origin_x: 0.0, - source_origin_y: 0.0, - source_width: 1.0, - source_height: 1.0, - pad2: 0, - } - } - - #[test] - fn backdrop_blur_pyramid_supports_six_passes_and_small_textures_limit_plans() { - let full_level_sizes = MetalRenderer::backdrop_blur_level_sizes_for(Size::new( - DevicePixels(4096), - DevicePixels(4096), - )); - assert_eq!(full_level_sizes.len(), MAX_BACKDROP_BLUR_LEVELS + 1); - assert_eq!( - MetalRenderer::backdrop_blur_plan_for_radius(240.0, &full_level_sizes).passes, - MAX_BACKDROP_BLUR_LEVELS - ); - - let small_level_sizes = MetalRenderer::backdrop_blur_level_sizes_for(Size::new( - DevicePixels(8), - DevicePixels(8), - )); - assert_eq!(small_level_sizes.len(), 3); - assert_eq!( - MetalRenderer::backdrop_blur_plan_for_radius(240.0, &small_level_sizes).passes, - 2 - ); - } - - #[test] - fn backdrop_blur_groups_require_equal_sample_distance() { - let level_sizes = MetalRenderer::backdrop_blur_level_sizes_for(Size::new( - DevicePixels(256), - DevicePixels(256), - )); - let blurs = [backdrop_blur(8.0), backdrop_blur(8.0), backdrop_blur(9.0)]; - let groups = MetalRenderer::backdrop_blur_plan_groups(&blurs, &level_sizes); - - assert_eq!(groups.len(), 2); - assert_eq!((groups[0].0, groups[0].1, groups[0].2.passes), (0, 2, 1)); - assert_eq!((groups[1].0, groups[1].1, groups[1].2.passes), (2, 3, 1)); - assert_ne!(groups[0].2.sample_distance, groups[1].2.sample_distance); - } - - #[test] - fn backdrop_blur_clusters_use_transitive_dilated_overlap_and_preserve_order() { - let viewport_size = Size::new(DevicePixels(200), DevicePixels(100)); - let blurs = [ - backdrop_blur_with_bounds(7, 0.0, 10.0, 1.0), - backdrop_blur_with_bounds(3, 30.0, 10.0, 1.0), - backdrop_blur_with_bounds(9, 11.0, 18.0, 1.0), - backdrop_blur_with_bounds(11, 150.0, 10.0, 1.0), - backdrop_blur_with_bounds(13, 300.0, 10.0, 1.0), - ]; - - let clusters = MetalRenderer::backdrop_blur_clusters(&blurs, viewport_size); - - assert_eq!(clusters.len(), 2); - assert_eq!( - clusters[0] - .iter() - .map(|blur| blur.order) - .collect::>(), - vec![7, 3, 9] - ); - assert_eq!( - clusters[1] - .iter() - .map(|blur| blur.order) - .collect::>(), - vec![11] - ); - } - - #[test] - fn backdrop_blur_texture_reuse_respects_cluster_max_bounds() { - let required = Size::new(DevicePixels(64), DevicePixels(64)); - let max_size = Size::new(DevicePixels(128), DevicePixels(128)); - - assert!(MetalRenderer::can_reuse_backdrop_texture( - Size::new(DevicePixels(128), DevicePixels(128)), - required, - max_size, - )); - assert!(!MetalRenderer::can_reuse_backdrop_texture( - Size::new(DevicePixels(512), DevicePixels(512)), - required, - max_size, - )); - assert!(!MetalRenderer::can_reuse_backdrop_texture( - Size::new(DevicePixels(32), DevicePixels(128)), - required, - max_size, - )); - } - #[test] fn backdrop_blur_snapshot_predicate_preserves_fast_path() { let positive_plan = BackdropBlurPlan { diff --git a/crates/gpui_wgpu/src/wgpu_renderer.rs b/crates/gpui_wgpu/src/wgpu_renderer.rs index e3a35e7..a76186f 100644 --- a/crates/gpui_wgpu/src/wgpu_renderer.rs +++ b/crates/gpui_wgpu/src/wgpu_renderer.rs @@ -4,7 +4,9 @@ use gpui::{ AtlasTextureId, BackdropBlur, BackdropBlurPlan, Background, Bounds, ContentMask, DevicePixels, GlobalElementId, GpuSpecs, MonochromeSprite, Path, Point, PolychromeSprite, PrimitiveBatch, Quad, RetainedLayer, RetainedLayerContentRevision, ScaledPixels, Scene, Shadow, Size, - SubpixelSprite, TransformationMatrix, Underline, get_gamma_correction_ratios, + SubpixelSprite, TransformationMatrix, Underline, backdrop_blur_clusters, + backdrop_blur_level_sizes_for, backdrop_blur_padding, backdrop_blur_plan_groups, + backdrop_source_bounds, can_reuse_backdrop_texture, get_gamma_correction_ratios, }; use log::warn; #[cfg(not(target_family = "wasm"))] @@ -307,6 +309,7 @@ pub struct WgpuResources { path_intermediate_size: Option>, backdrop_texture: Option, backdrop_view: Option, + backdrop_blur_input_view: Option, backdrop_size: Option>, backdrop_blur_level_sizes: Vec>, backdrop_blur_downsample_textures: Vec, @@ -320,6 +323,7 @@ impl WgpuResources { fn discard_backdrop_resources(&mut self) { self.backdrop_texture = None; self.backdrop_view = None; + self.backdrop_blur_input_view = None; self.backdrop_size = None; self.backdrop_blur_level_sizes.clear(); self.backdrop_blur_downsample_textures.clear(); @@ -731,6 +735,7 @@ impl WgpuRenderer { path_intermediate_size: None, backdrop_texture: None, backdrop_view: None, + backdrop_blur_input_view: None, backdrop_size: None, backdrop_blur_level_sizes: Vec::new(), backdrop_blur_downsample_textures: Vec::new(), @@ -1204,7 +1209,9 @@ impl WgpuRenderer { format: wgpu::TextureFormat, width: u32, height: u32, - ) -> (wgpu::Texture, wgpu::TextureView) { + ) -> (wgpu::Texture, wgpu::TextureView, wgpu::TextureView) { + let linear_format = backdrop_blur_format(format); + let view_formats = (linear_format != format).then_some([linear_format]); let texture = device.create_texture(&wgpu::TextureDescriptor { label: Some("backdrop_texture"), size: wgpu::Extent3d { @@ -1217,10 +1224,14 @@ impl WgpuRenderer { dimension: wgpu::TextureDimension::D2, format, usage: wgpu::TextureUsages::COPY_DST | wgpu::TextureUsages::TEXTURE_BINDING, - view_formats: &[], + view_formats: view_formats.as_ref().map_or(&[], |formats| formats), }); let view = texture.create_view(&wgpu::TextureViewDescriptor::default()); - (texture, view) + let blur_input_view = texture.create_view(&wgpu::TextureViewDescriptor { + format: Some(linear_format), + ..Default::default() + }); + (texture, view, blur_input_view) } fn create_backdrop_blur_texture( @@ -1401,7 +1412,7 @@ impl WgpuRenderer { if let Some(current_size) = self.backdrop_size && self.backdrop_texture.is_some() { - if Self::can_reuse_backdrop_texture(current_size, size, max_size) { + if can_reuse_backdrop_texture(current_size, size) { return Some(current_size); } return self.create_backdrop_resources(Size { @@ -1413,17 +1424,6 @@ impl WgpuRenderer { self.create_backdrop_resources(size) } - fn can_reuse_backdrop_texture( - current_size: Size, - required_size: Size, - max_size: Size, - ) -> bool { - current_size.width >= required_size.width - && current_size.height >= required_size.height - && current_size.width <= max_size.width - && current_size.height <= max_size.height - } - fn quantize_backdrop_texture_size( size: Size, max_size: Size, @@ -1454,13 +1454,14 @@ impl WgpuRenderer { return None; } - let (backdrop_texture, backdrop_view) = Self::create_backdrop_texture( - &self.device, - self.surface_config.format, - size.width.0 as u32, - size.height.0 as u32, - ); - let level_sizes = Self::backdrop_blur_level_sizes_for(size); + let (backdrop_texture, backdrop_view, backdrop_blur_input_view) = + Self::create_backdrop_texture( + &self.device, + self.surface_config.format, + size.width.0 as u32, + size.height.0 as u32, + ); + let level_sizes = backdrop_blur_level_sizes_for(size); let blur_format = backdrop_blur_format(self.surface_config.format); let mut downsample_textures = Vec::with_capacity(level_sizes.len().saturating_sub(1)); let mut downsample_views = Vec::with_capacity(level_sizes.len().saturating_sub(1)); @@ -1488,6 +1489,7 @@ impl WgpuRenderer { } self.backdrop_texture = Some(backdrop_texture); self.backdrop_view = Some(backdrop_view); + self.backdrop_blur_input_view = Some(backdrop_blur_input_view); self.backdrop_size = Some(size); self.backdrop_blur_level_sizes = level_sizes; self.backdrop_blur_downsample_textures = downsample_textures; @@ -1585,7 +1587,7 @@ impl WgpuRenderer { if !scene .backdrop_blurs .iter() - .any(|blur| Self::backdrop_source_bounds(blur, viewport_size).is_some()) + .any(|blur| backdrop_source_bounds(blur, viewport_size).is_some()) { self.discard_backdrop_resources(); } @@ -2137,7 +2139,7 @@ impl WgpuRenderer { if blurs.is_empty() || !target_supports_copy_src { continue; } - let blur_clusters = Self::backdrop_blur_clusters(blurs, target_size); + let blur_clusters = backdrop_blur_clusters(blurs, target_size); if blur_clusters.is_empty() { continue; } @@ -2155,15 +2157,19 @@ impl WgpuRenderer { ) else { continue; }; - scratch_bounds.texture_size = texture_size; + scratch_bounds = Self::fit_backdrop_scratch_bounds( + scratch_bounds, + texture_size, + target_size, + ); let Some(backdrop_texture) = self.backdrop_texture.as_ref() else { continue; }; let prepared_blurs = Self::prepare_backdrop_blurs(&blurs, scratch_bounds); - let plan_groups = Self::backdrop_blur_plan_groups( + let plan_groups = backdrop_blur_plan_groups( &blurs, - &self.backdrop_blur_level_sizes, + self.backdrop_blur_level_sizes.len().saturating_sub(1), ); drop(pass); encoder.copy_texture_to_texture( @@ -2492,7 +2498,7 @@ impl WgpuRenderer { for level in 0..plan.passes { let input_view = if level == 0 { - let Some(view) = self.backdrop_view.as_ref() else { + let Some(view) = self.backdrop_blur_input_view.as_ref() else { return false; }; view @@ -2774,12 +2780,12 @@ impl WgpuRenderer { let mut bounds = blurs .first()? .bounds - .dilate(Self::backdrop_blur_padding(blurs.first()?.blur_radius.0)); + .dilate(backdrop_blur_padding(blurs.first()?.blur_radius.0)); for blur in blurs.iter().skip(1) { bounds = bounds.union( &blur .bounds - .dilate(Self::backdrop_blur_padding(blur.blur_radius.0)), + .dilate(backdrop_blur_padding(blur.blur_radius.0)), ); } @@ -2810,75 +2816,6 @@ impl WgpuRenderer { }) } - fn backdrop_source_bounds( - blur: &BackdropBlur, - viewport_size: Size, - ) -> Option> { - let viewport_bounds = Bounds { - origin: Point { - x: ScaledPixels(0.0), - y: ScaledPixels(0.0), - }, - size: Size { - width: ScaledPixels::from(viewport_size.width), - height: ScaledPixels::from(viewport_size.height), - }, - }; - let bounds = blur - .bounds - .dilate(Self::backdrop_blur_padding(blur.blur_radius.0)) - .intersect(&viewport_bounds); - if bounds.is_empty() { - return None; - } - - let origin = bounds.origin.map(|component| component.floor()); - let bottom_right = bounds.bottom_right().map(|component| component.ceil()); - Some(Bounds::from_corners(origin, bottom_right)) - } - - fn backdrop_blur_clusters( - blurs: &[BackdropBlur], - viewport_size: Size, - ) -> Vec> { - let mut clusters: Vec<(Bounds, Vec<(usize, BackdropBlur)>)> = Vec::new(); - for (blur_ix, blur) in blurs.iter().enumerate() { - let Some(mut cluster_bounds) = Self::backdrop_source_bounds(blur, viewport_size) else { - continue; - }; - let mut cluster_blurs = vec![(blur_ix, blur.clone())]; - - loop { - let overlapping_cluster_ixs = clusters - .iter() - .enumerate() - .filter_map(|(ix, (bounds, _))| { - bounds.intersects(&cluster_bounds).then_some(ix) - }) - .collect::>(); - if overlapping_cluster_ixs.is_empty() { - break; - } - - for ix in overlapping_cluster_ixs.into_iter().rev() { - let (bounds, mut blurs) = clusters.remove(ix); - cluster_bounds = cluster_bounds.union(&bounds); - cluster_blurs.append(&mut blurs); - } - } - - clusters.push((cluster_bounds, cluster_blurs)); - } - - clusters - .into_iter() - .map(|(_, mut blurs)| { - blurs.sort_by_key(|(ix, _)| *ix); - blurs.into_iter().map(|(_, blur)| blur).collect() - }) - .collect() - } - fn max_backdrop_texture_size( scratch_bounds: BackdropScratchBounds, viewport_size: Size, @@ -2893,65 +2830,30 @@ impl WgpuRenderer { } } - fn backdrop_blur_padding(radius: f32) -> ScaledPixels { - ScaledPixels(BackdropBlurPlan::padding(radius)) - } - - fn backdrop_blur_level_sizes_for(size: Size) -> Vec> { - let mut level_sizes = Vec::new(); - if size.width.0 <= 0 || size.height.0 <= 0 { - return level_sizes; - } - - level_sizes.push(size); - let mut level_size = size; - for _ in 0..MAX_BACKDROP_BLUR_LEVELS { - let next_width = level_size.width.0 / 2; - let next_height = level_size.height.0 / 2; - if next_width < 2 || next_height < 2 { - break; - } - level_size = Size { - width: DevicePixels(next_width), - height: DevicePixels(next_height), - }; - level_sizes.push(level_size); - } - level_sizes - } - - fn backdrop_blur_plan_for_radius( - radius: f32, - level_sizes: &[Size], - ) -> BackdropBlurPlan { - BackdropBlurPlan::for_radius( - radius, - level_sizes - .len() - .saturating_sub(1) - .min(MAX_BACKDROP_BLUR_LEVELS), - ) - } - - fn backdrop_blur_plan_groups( - blurs: &[BackdropBlur], - level_sizes: &[Size], - ) -> Vec<(usize, usize, BackdropBlurPlan)> { - let mut groups = Vec::new(); - let mut start = 0; - while start < blurs.len() { - let plan = Self::backdrop_blur_plan_for_radius(blurs[start].blur_radius.0, level_sizes); - let mut end = start + 1; - while end < blurs.len() - && Self::backdrop_blur_plan_for_radius(blurs[end].blur_radius.0, level_sizes) - == plan - { - end += 1; - } - groups.push((start, end, plan)); - start = end; - } - groups + fn fit_backdrop_scratch_bounds( + mut scratch_bounds: BackdropScratchBounds, + texture_size: Size, + viewport_size: Size, + ) -> BackdropScratchBounds { + let texture_size = Size { + width: texture_size.width.min(viewport_size.width), + height: texture_size.height.min(viewport_size.height), + }; + let max_origin_x = (viewport_size.width.0 - texture_size.width.0).max(0) as f32; + let max_origin_y = (viewport_size.height.0 - texture_size.height.0).max(0) as f32; + let origin = Point { + x: ScaledPixels(scratch_bounds.bounds.origin.x.0.clamp(0.0, max_origin_x)), + y: ScaledPixels(scratch_bounds.bounds.origin.y.0.clamp(0.0, max_origin_y)), + }; + scratch_bounds.bounds = Bounds { + origin, + size: Size { + width: ScaledPixels::from(texture_size.width), + height: ScaledPixels::from(texture_size.height), + }, + }; + scratch_bounds.texture_size = texture_size; + scratch_bounds } fn prepare_backdrop_blurs( @@ -3322,98 +3224,6 @@ mod tests { (scene, layer) } - fn backdrop_blur(radius: f32) -> BackdropBlur { - backdrop_blur_with_bounds(0, 0.0, 100.0, radius) - } - - fn backdrop_blur_with_bounds(order: u32, x: f32, width: f32, radius: f32) -> BackdropBlur { - let bounds = Bounds::new( - Point::new(ScaledPixels(x), ScaledPixels(0.0)), - Size::new(ScaledPixels(width), ScaledPixels(10.0)), - ); - BackdropBlur { - order, - pad: 0, - bounds, - content_mask: ContentMask::new(bounds), - corner_radii: Corners::default(), - blur_radius: ScaledPixels(radius), - source_origin_x: 0.0, - source_origin_y: 0.0, - source_width: 1.0, - source_height: 1.0, - pad2: 0, - } - } - - #[test] - fn backdrop_blur_pyramid_supports_six_passes_and_small_textures_limit_plans() { - let full_level_sizes = WgpuRenderer::backdrop_blur_level_sizes_for(Size::new( - DevicePixels(4096), - DevicePixels(4096), - )); - assert_eq!(full_level_sizes.len(), MAX_BACKDROP_BLUR_LEVELS + 1); - assert_eq!( - WgpuRenderer::backdrop_blur_plan_for_radius(240.0, &full_level_sizes).passes, - MAX_BACKDROP_BLUR_LEVELS - ); - - let small_level_sizes = WgpuRenderer::backdrop_blur_level_sizes_for(Size::new( - DevicePixels(8), - DevicePixels(8), - )); - assert_eq!(small_level_sizes.len(), 3); - assert_eq!( - WgpuRenderer::backdrop_blur_plan_for_radius(240.0, &small_level_sizes).passes, - 2 - ); - } - - #[test] - fn backdrop_blur_groups_require_equal_sample_distance() { - let level_sizes = WgpuRenderer::backdrop_blur_level_sizes_for(Size::new( - DevicePixels(256), - DevicePixels(256), - )); - let blurs = [backdrop_blur(8.0), backdrop_blur(8.0), backdrop_blur(9.0)]; - let groups = WgpuRenderer::backdrop_blur_plan_groups(&blurs, &level_sizes); - - assert_eq!(groups.len(), 2); - assert_eq!((groups[0].0, groups[0].1, groups[0].2.passes), (0, 2, 1)); - assert_eq!((groups[1].0, groups[1].1, groups[1].2.passes), (2, 3, 1)); - assert_ne!(groups[0].2.sample_distance, groups[1].2.sample_distance); - } - - #[test] - fn backdrop_blur_clusters_use_transitive_dilated_overlap_and_preserve_order() { - let viewport_size = Size::new(DevicePixels(200), DevicePixels(100)); - let blurs = [ - backdrop_blur_with_bounds(7, 0.0, 10.0, 1.0), - backdrop_blur_with_bounds(3, 30.0, 10.0, 1.0), - backdrop_blur_with_bounds(9, 11.0, 18.0, 1.0), - backdrop_blur_with_bounds(11, 150.0, 10.0, 1.0), - backdrop_blur_with_bounds(13, 300.0, 10.0, 1.0), - ]; - - let clusters = WgpuRenderer::backdrop_blur_clusters(&blurs, viewport_size); - - assert_eq!(clusters.len(), 2); - assert_eq!( - clusters[0] - .iter() - .map(|blur| blur.order) - .collect::>(), - vec![7, 3, 9] - ); - assert_eq!( - clusters[1] - .iter() - .map(|blur| blur.order) - .collect::>(), - vec![11] - ); - } - #[test] fn backdrop_blur_pass_params_match_wgsl_storage_layout() { assert_eq!(std::mem::size_of::(), 16); @@ -3444,28 +3254,6 @@ mod tests { ); } - #[test] - fn backdrop_blur_texture_reuse_respects_cluster_max_bounds() { - let required = Size::new(DevicePixels(64), DevicePixels(64)); - let max_size = Size::new(DevicePixels(128), DevicePixels(128)); - - assert!(WgpuRenderer::can_reuse_backdrop_texture( - Size::new(DevicePixels(128), DevicePixels(128)), - required, - max_size, - )); - assert!(!WgpuRenderer::can_reuse_backdrop_texture( - Size::new(DevicePixels(512), DevicePixels(512)), - required, - max_size, - )); - assert!(!WgpuRenderer::can_reuse_backdrop_texture( - Size::new(DevicePixels(32), DevicePixels(128)), - required, - max_size, - )); - } - #[test] fn shaders_parse_and_validate() { let module = diff --git a/crates/gpui_windows/src/directx_renderer.rs b/crates/gpui_windows/src/directx_renderer.rs index 2790054..93c3667 100644 --- a/crates/gpui_windows/src/directx_renderer.rs +++ b/crates/gpui_windows/src/directx_renderer.rs @@ -542,10 +542,14 @@ impl DirectXRenderer { // and so likely do not have the textures anymore that are required for drawing return Ok(()); } + let viewport_size = size( + DevicePixels(self.width as i32), + DevicePixels(self.height as i32), + ); if !scene .backdrop_blurs .iter() - .any(|blur| Self::backdrop_source_bounds(blur, self.width, self.height).is_some()) + .any(|blur| backdrop_source_bounds(blur, viewport_size).is_some()) { if let Some(resources) = self.resources.as_mut() { resources.discard_backdrop_resources(); @@ -569,7 +573,11 @@ impl DirectXRenderer { if blurs.is_empty() { Ok(()) } else { - for blurs in Self::backdrop_blur_clusters(blurs, self.width, self.height) { + let viewport_size = size( + DevicePixels(self.width as i32), + DevicePixels(self.height as i32), + ); + for blurs in backdrop_blur_clusters(blurs, viewport_size) { let Some(mut scratch_bounds) = Self::backdrop_scratch_bounds(&blurs, self.width, self.height) else { @@ -591,7 +599,11 @@ impl DirectXRenderer { ), )? { - scratch_bounds.texture_size = texture_size; + scratch_bounds = Self::fit_backdrop_scratch_bounds( + scratch_bounds, + texture_size, + viewport_size, + ); } } let prepared_blurs = @@ -600,20 +612,12 @@ impl DirectXRenderer { let max_backdrop_blur_levels = self.max_backdrop_blur_levels(); let mut current_plan = None; let mut current_blur_srv = None; - let mut start = 0; - while start < blurs.len() { - let plan = Self::backdrop_blur_plan_for_radius( - blurs[start].blur_radius.0, - max_backdrop_blur_levels, - ); - let mut end = start + 1; - while end < blurs.len() - && Self::backdrop_blur_plan_for_radius( - blurs[end].blur_radius.0, - max_backdrop_blur_levels, - ) == plan - { - end += 1; + for (start, end, plan) in backdrop_blur_plan_groups( + &blurs, + max_backdrop_blur_levels, + ) { + if plan.passes == 0 { + continue; } if current_plan != Some(plan) { current_blur_srv = @@ -624,7 +628,6 @@ impl DirectXRenderer { &prepared_blurs[start..end], ¤t_blur_srv, )?; - start = end; } } Ok(()) @@ -1028,12 +1031,12 @@ impl DirectXRenderer { let mut bounds = blurs .first()? .bounds - .dilate(Self::backdrop_blur_padding(blurs.first()?.blur_radius.0)); + .dilate(backdrop_blur_padding(blurs.first()?.blur_radius.0)); for blur in blurs.iter().skip(1) { bounds = bounds.union( &blur .bounds - .dilate(Self::backdrop_blur_padding(blur.blur_radius.0)), + .dilate(backdrop_blur_padding(blur.blur_radius.0)), ); } @@ -1058,71 +1061,6 @@ impl DirectXRenderer { }) } - fn backdrop_source_bounds( - blur: &BackdropBlur, - width: u32, - height: u32, - ) -> Option> { - let viewport_bounds = Bounds { - origin: point(ScaledPixels(0.0), ScaledPixels(0.0)), - size: size(ScaledPixels(width as f32), ScaledPixels(height as f32)), - }; - let bounds = blur - .bounds - .dilate(Self::backdrop_blur_padding(blur.blur_radius.0)) - .intersect(&viewport_bounds); - if bounds.is_empty() { - return None; - } - - let origin = bounds.origin.map(|component| component.floor()); - let bottom_right = bounds.bottom_right().map(|component| component.ceil()); - Some(Bounds::from_corners(origin, bottom_right)) - } - - fn backdrop_blur_clusters( - blurs: &[BackdropBlur], - width: u32, - height: u32, - ) -> Vec> { - let mut clusters: Vec<(Bounds, Vec<(usize, BackdropBlur)>)> = Vec::new(); - for (blur_ix, blur) in blurs.iter().enumerate() { - let Some(mut cluster_bounds) = Self::backdrop_source_bounds(blur, width, height) else { - continue; - }; - let mut cluster_blurs = vec![(blur_ix, blur.clone())]; - - loop { - let overlapping_cluster_ixs = clusters - .iter() - .enumerate() - .filter_map(|(ix, (bounds, _))| { - bounds.intersects(&cluster_bounds).then_some(ix) - }) - .collect::>(); - if overlapping_cluster_ixs.is_empty() { - break; - } - - for ix in overlapping_cluster_ixs.into_iter().rev() { - let (bounds, mut blurs) = clusters.remove(ix); - cluster_bounds = cluster_bounds.union(&bounds); - cluster_blurs.append(&mut blurs); - } - } - - clusters.push((cluster_bounds, cluster_blurs)); - } - - clusters - .into_iter() - .map(|(_, mut blurs)| { - blurs.sort_by_key(|(ix, _)| *ix); - blurs.into_iter().map(|(_, blur)| blur).collect() - }) - .collect() - } - fn max_backdrop_texture_size( scratch_bounds: BackdropScratchBounds, width: u32, @@ -1134,8 +1072,30 @@ impl DirectXRenderer { } } - fn backdrop_blur_padding(radius: f32) -> ScaledPixels { - ScaledPixels(BackdropBlurPlan::padding(radius)) + fn fit_backdrop_scratch_bounds( + mut scratch_bounds: BackdropScratchBounds, + texture_size: Size, + viewport_size: Size, + ) -> BackdropScratchBounds { + let texture_size = size( + texture_size.width.min(viewport_size.width), + texture_size.height.min(viewport_size.height), + ); + let max_origin_x = (viewport_size.width.0 - texture_size.width.0).max(0) as f32; + let max_origin_y = (viewport_size.height.0 - texture_size.height.0).max(0) as f32; + let origin = point( + ScaledPixels(scratch_bounds.bounds.origin.x.0.clamp(0.0, max_origin_x)), + ScaledPixels(scratch_bounds.bounds.origin.y.0.clamp(0.0, max_origin_y)), + ); + scratch_bounds.bounds = Bounds { + origin, + size: size( + ScaledPixels::from(texture_size.width), + ScaledPixels::from(texture_size.height), + ), + }; + scratch_bounds.texture_size = texture_size; + scratch_bounds } fn prepare_backdrop_blurs( @@ -1164,10 +1124,6 @@ impl DirectXRenderer { .min(MAX_BACKDROP_BLUR_LEVELS) } - fn backdrop_blur_plan_for_radius(radius: f32, max_levels: usize) -> BackdropBlurPlan { - BackdropBlurPlan::for_radius(radius, max_levels.min(MAX_BACKDROP_BLUR_LEVELS)) - } - fn draw_backdrop_blurs( &mut self, blurs: &[BackdropBlur], @@ -1809,11 +1765,7 @@ impl DirectXResources { if let Some(current_size) = self.backdrop_size && self.backdrop_texture.is_some() { - if current_size.width >= size.width - && current_size.height >= size.height - && current_size.width <= max_size.width - && current_size.height <= max_size.height - { + if can_reuse_backdrop_texture(current_size, size) { return Ok(Some(current_size)); } return self.create_backdrop_resources( @@ -2616,27 +2568,6 @@ fn create_backdrop_texture_and_srv( Ok((texture, srv)) } -fn backdrop_blur_level_sizes(width: u32, height: u32) -> Vec<(u32, u32)> { - let mut levels = Vec::new(); - if width == 0 || height == 0 { - return levels; - } - levels.push((width, height)); - let mut current_width = width; - let mut current_height = height; - for _ in 0..MAX_BACKDROP_BLUR_LEVELS { - let next_width = current_width / 2; - let next_height = current_height / 2; - if next_width < 2 || next_height < 2 { - break; - } - current_width = next_width; - current_height = next_height; - levels.push((current_width, current_height)); - } - levels -} - #[inline] fn create_backdrop_blur_texture_and_views( device: &ID3D11Device, @@ -2683,7 +2614,13 @@ fn create_backdrop_blur_resources( width: u32, height: u32, ) -> Result { - let level_sizes = backdrop_blur_level_sizes(width, height); + let level_sizes = backdrop_blur_level_sizes_for(size( + DevicePixels(width as i32), + DevicePixels(height as i32), + )) + .into_iter() + .map(|level_size| (level_size.width.0 as u32, level_size.height.0 as u32)) + .collect::>(); let mut downsample_textures = Vec::new(); let mut downsample_views = Vec::new(); let mut downsample_srvs = Vec::new(); @@ -3387,35 +3324,13 @@ mod amd { #[cfg(test)] mod tests { - use gpui::{ - BackdropBlur, Bounds, ContentMask, Corners, Point, ScaledPixels, Size, point, size, - }; + use gpui::{Bounds, ContentMask, Corners, ScaledPixels, point, size}; use super::{ - BackdropBlurParams, DirectXRenderer, GlobalParams, RetainedLayerClip, - backdrop_blur_level_sizes, retained_layer_clip, rounded_backdrop_rebuild_requested, + BackdropBlurParams, GlobalParams, RetainedLayerClip, retained_layer_clip, + rounded_backdrop_rebuild_requested, }; - fn backdrop_blur_with_bounds(order: u32, x: f32, width: f32, radius: f32) -> BackdropBlur { - let bounds = Bounds::new( - Point::new(ScaledPixels(x), ScaledPixels(0.0)), - Size::new(ScaledPixels(width), ScaledPixels(10.0)), - ); - BackdropBlur { - order, - pad: 0, - bounds, - content_mask: ContentMask::new(bounds), - corner_radii: Corners::default(), - blur_radius: ScaledPixels(radius), - source_origin_x: 0.0, - source_origin_y: 0.0, - source_width: 1.0, - source_height: 1.0, - pad2: 0, - } - } - #[test] fn global_params_preserve_hlsl_constant_buffer_alignment() { assert_eq!(std::mem::size_of::(), 48); @@ -3426,46 +3341,6 @@ mod tests { assert_eq!(std::mem::size_of::(), 16); } - #[test] - fn backdrop_blur_pyramid_caps_levels_and_groups_by_full_plan() { - assert_eq!(backdrop_blur_level_sizes(4_096, 4_096).len(), 7); - assert_eq!(backdrop_blur_level_sizes(3, 3), vec![(3, 3)]); - - let first = DirectXRenderer::backdrop_blur_plan_for_radius(8.0, 6); - let second = DirectXRenderer::backdrop_blur_plan_for_radius(9.0, 6); - assert_eq!(first.passes, second.passes); - assert_ne!(first, second); - } - - #[test] - fn backdrop_blur_clusters_use_transitive_dilated_overlap_and_preserve_order() { - let blurs = [ - backdrop_blur_with_bounds(7, 0.0, 10.0, 1.0), - backdrop_blur_with_bounds(3, 30.0, 10.0, 1.0), - backdrop_blur_with_bounds(9, 11.0, 18.0, 1.0), - backdrop_blur_with_bounds(11, 150.0, 10.0, 1.0), - backdrop_blur_with_bounds(13, 300.0, 10.0, 1.0), - ]; - - let clusters = DirectXRenderer::backdrop_blur_clusters(&blurs, 200, 100); - - assert_eq!(clusters.len(), 2); - assert_eq!( - clusters[0] - .iter() - .map(|blur| blur.order) - .collect::>(), - vec![7, 3, 9] - ); - assert_eq!( - clusters[1] - .iter() - .map(|blur| blur.order) - .collect::>(), - vec![11] - ); - } - #[test] fn rounded_backdrop_rebuild_uses_retained_configuration() { assert!(!rounded_backdrop_rebuild_requested(None)); From 0703fd41bd6fca8f102ec3d06620c3bae0ef28c4 Mon Sep 17 00:00:00 2001 From: Aditya Sharma Date: Mon, 27 Jul 2026 22:06:18 -0700 Subject: [PATCH 3/4] fix(rendering): exclude no-op backdrop blurs Keep invalid radii out of source clustering and centralize scratch-bound preparation so every renderer preserves the same viewport and reuse rules. --- crates/gpui/src/scene.rs | 168 ++++++++++++++++++++ crates/gpui_macos/src/metal_renderer.rs | 123 ++------------ crates/gpui_wgpu/src/wgpu_renderer.rs | 120 +------------- crates/gpui_windows/src/directx_renderer.rs | 111 +------------ 4 files changed, 190 insertions(+), 332 deletions(-) diff --git a/crates/gpui/src/scene.rs b/crates/gpui/src/scene.rs index 3c411e4..5269093 100644 --- a/crates/gpui/src/scene.rs +++ b/crates/gpui/src/scene.rs @@ -898,12 +898,26 @@ pub fn can_reuse_backdrop_texture( current_size.width >= required_size.width && current_size.height >= required_size.height } +/// Bounds and texture dimensions for backdrop blur scratch rendering. +#[doc(hidden)] +#[derive(Clone, Copy)] +pub struct BackdropScratchBounds { + /// Viewport-space source bounds. + pub bounds: Bounds, + /// Scratch texture dimensions. + pub texture_size: Size, +} + /// Returns the viewport-clipped source bounds required for one backdrop blur. #[doc(hidden)] pub fn backdrop_source_bounds( blur: &BackdropBlur, viewport_size: Size, ) -> Option> { + if !blur.blur_radius.0.is_finite() || blur.blur_radius.0 <= 0.0 { + return None; + } + let viewport_bounds = Bounds { origin: point(ScaledPixels(0.0), ScaledPixels(0.0)), size: Size { @@ -959,6 +973,111 @@ pub fn backdrop_blur_clusters( clusters.into_iter().map(|(_, blurs)| blurs).collect() } +/// Returns viewport-clipped scratch bounds for a backdrop blur cluster. +#[doc(hidden)] +pub fn backdrop_scratch_bounds( + blurs: &[BackdropBlur], + viewport_size: Size, +) -> Option { + let mut bounds = blurs + .first()? + .bounds + .dilate(backdrop_blur_padding(blurs.first()?.blur_radius.0)); + for blur in blurs.iter().skip(1) { + bounds = bounds.union( + &blur + .bounds + .dilate(backdrop_blur_padding(blur.blur_radius.0)), + ); + } + + let viewport_bounds = Bounds { + origin: point(ScaledPixels(0.0), ScaledPixels(0.0)), + size: Size { + width: ScaledPixels::from(viewport_size.width), + height: ScaledPixels::from(viewport_size.height), + }, + }; + bounds = bounds.intersect(&viewport_bounds); + if bounds.is_empty() { + return None; + } + + let origin = bounds.origin.map(|component| component.floor()); + let bottom_right = bounds.bottom_right().map(|component| component.ceil()); + let bounds = Bounds::from_corners(origin, bottom_right); + Some(BackdropScratchBounds { + texture_size: Size { + width: DevicePixels::from(bounds.size.width), + height: DevicePixels::from(bounds.size.height), + }, + bounds, + }) +} + +/// Returns the largest texture that fits from the scratch origin to the viewport edge. +#[doc(hidden)] +pub fn max_backdrop_texture_size( + scratch_bounds: BackdropScratchBounds, + viewport_size: Size, +) -> Size { + Size { + width: DevicePixels( + (viewport_size.width.0 - scratch_bounds.bounds.origin.x.0 as i32).max(0), + ), + height: DevicePixels( + (viewport_size.height.0 - scratch_bounds.bounds.origin.y.0 as i32).max(0), + ), + } +} + +/// Fits an allocated backdrop texture within the viewport. +#[doc(hidden)] +pub fn fit_backdrop_scratch_bounds( + mut scratch_bounds: BackdropScratchBounds, + texture_size: Size, + viewport_size: Size, +) -> BackdropScratchBounds { + let texture_size = Size { + width: texture_size.width.min(viewport_size.width), + height: texture_size.height.min(viewport_size.height), + }; + let max_origin_x = (viewport_size.width.0 - texture_size.width.0).max(0) as f32; + let max_origin_y = (viewport_size.height.0 - texture_size.height.0).max(0) as f32; + let origin = Point { + x: ScaledPixels(scratch_bounds.bounds.origin.x.0.clamp(0.0, max_origin_x)), + y: ScaledPixels(scratch_bounds.bounds.origin.y.0.clamp(0.0, max_origin_y)), + }; + scratch_bounds.bounds = Bounds { + origin, + size: Size { + width: ScaledPixels::from(texture_size.width), + height: ScaledPixels::from(texture_size.height), + }, + }; + scratch_bounds.texture_size = texture_size; + scratch_bounds +} + +/// Prepares backdrop blur instances to sample from shared scratch bounds. +#[doc(hidden)] +pub fn prepare_backdrop_blurs( + blurs: &[BackdropBlur], + scratch_bounds: BackdropScratchBounds, +) -> Vec { + blurs + .iter() + .cloned() + .map(|mut blur| { + blur.source_origin_x = scratch_bounds.bounds.origin.x.0; + blur.source_origin_y = scratch_bounds.bounds.origin.y.0; + blur.source_width = scratch_bounds.texture_size.width.0 as f32; + blur.source_height = scratch_bounds.texture_size.height.0 as f32; + blur + }) + .collect() +} + #[derive(Debug, Clone)] #[repr(C)] #[expect(missing_docs)] @@ -1573,6 +1692,55 @@ mod tests { )); } + #[test] + fn backdrop_blur_clusters_exclude_noop_radii() { + let blurs = [ + test_backdrop_blur_with_bounds(1, 0.0, 10.0, 0.0), + test_backdrop_blur_with_bounds(2, 0.0, 10.0, -1.0), + test_backdrop_blur_with_bounds(3, 0.0, 10.0, f32::NAN), + ]; + let viewport_size = size(DevicePixels(200), DevicePixels(100)); + + assert!( + blurs + .iter() + .all(|blur| backdrop_source_bounds(blur, viewport_size).is_none()) + ); + assert!(backdrop_blur_clusters(&blurs, viewport_size).is_empty()); + } + + #[test] + fn backdrop_blur_scratch_helpers_fit_reused_textures() { + let blurs = [test_backdrop_blur_with_bounds(1, 80.0, 10.0, 1.0)]; + let viewport_size = size(DevicePixels(100), DevicePixels(100)); + let scratch_bounds = backdrop_scratch_bounds(&blurs, viewport_size).unwrap(); + + assert_eq!( + max_backdrop_texture_size(scratch_bounds, viewport_size), + size(DevicePixels(26), DevicePixels(100)) + ); + + let fitted = fit_backdrop_scratch_bounds( + scratch_bounds, + size(DevicePixels(64), DevicePixels(64)), + viewport_size, + ); + assert_eq!( + fitted.bounds.origin, + point(ScaledPixels(36.0), ScaledPixels(0.0)) + ); + assert_eq!( + fitted.texture_size, + size(DevicePixels(64), DevicePixels(64)) + ); + + let prepared = prepare_backdrop_blurs(&blurs, fitted); + assert_eq!(prepared[0].source_origin_x, 36.0); + assert_eq!(prepared[0].source_origin_y, 0.0); + assert_eq!(prepared[0].source_width, 64.0); + assert_eq!(prepared[0].source_height, 64.0); + } + #[test] fn backdrop_blur_clusters_preserve_interleaved_source_order() { let blurs = [ diff --git a/crates/gpui_macos/src/metal_renderer.rs b/crates/gpui_macos/src/metal_renderer.rs index ef6fdb3..4cd87a7 100644 --- a/crates/gpui_macos/src/metal_renderer.rs +++ b/crates/gpui_macos/src/metal_renderer.rs @@ -7,11 +7,13 @@ use cocoa::{ quartzcore::AutoresizingMask, }; use gpui::{ - AtlasTextureId, BackdropBlur, BackdropBlurPlan, Background, Bounds, ContentMask, DevicePixels, - GlobalElementId, MonochromeSprite, PaintSurface, Path, Point, PolychromeSprite, PrimitiveBatch, - Quad, RetainedLayer, RetainedLayerContentRevision, ScaledPixels, Scene, Shadow, Size, Surface, - TransformationMatrix, Underline, backdrop_blur_clusters, backdrop_blur_level_sizes_for, - backdrop_blur_padding, backdrop_blur_plan_groups, can_reuse_backdrop_texture, point, size, + AtlasTextureId, BackdropBlur, BackdropBlurPlan, BackdropScratchBounds, Background, Bounds, + ContentMask, DevicePixels, GlobalElementId, MonochromeSprite, PaintSurface, Path, Point, + PolychromeSprite, PrimitiveBatch, Quad, RetainedLayer, RetainedLayerContentRevision, + ScaledPixels, Scene, Shadow, Size, Surface, TransformationMatrix, Underline, + backdrop_blur_clusters, backdrop_blur_level_sizes_for, backdrop_blur_plan_groups, + backdrop_scratch_bounds, can_reuse_backdrop_texture, fit_backdrop_scratch_bounds, + max_backdrop_texture_size, point, prepare_backdrop_blurs, size, }; #[cfg(any(test, feature = "test-support"))] use image::RgbaImage; @@ -229,12 +231,6 @@ struct PathScratchBounds { texture_size: Size, } -#[derive(Clone, Copy)] -struct BackdropScratchBounds { - bounds: Bounds, - texture_size: Size, -} - #[repr(C)] struct BackdropBlurParams { input_origin: [f32; 2], @@ -1410,23 +1406,23 @@ impl MetalRenderer { let mut ok = true; for blurs in backdrop_blur_clusters(blurs, viewport_size) { let Some(mut scratch_bounds) = - Self::backdrop_scratch_bounds(&blurs, viewport_size) + backdrop_scratch_bounds(&blurs, viewport_size) else { continue; }; let Some(texture_size) = self.ensure_backdrop_textures( scratch_bounds.texture_size, - Self::max_backdrop_texture_size(scratch_bounds, viewport_size), + max_backdrop_texture_size(scratch_bounds, viewport_size), ) else { continue; }; - scratch_bounds = Self::fit_backdrop_scratch_bounds( + scratch_bounds = fit_backdrop_scratch_bounds( scratch_bounds, texture_size, viewport_size, ); let level_sizes = self.backdrop_blur_level_sizes.clone(); - let prepared_blurs = Self::prepare_backdrop_blurs(&blurs, scratch_bounds); + let prepared_blurs = prepare_backdrop_blurs(&blurs, scratch_bounds); let plan_groups = backdrop_blur_plan_groups(&blurs, level_sizes.len().saturating_sub(1)); @@ -1697,103 +1693,6 @@ impl MetalRenderer { command_encoder.end_encoding(); } - fn backdrop_scratch_bounds( - blurs: &[BackdropBlur], - viewport_size: Size, - ) -> Option { - let mut bounds = blurs - .first()? - .bounds - .dilate(backdrop_blur_padding(blurs.first()?.blur_radius.0)); - for blur in blurs.iter().skip(1) { - bounds = bounds.union( - &blur - .bounds - .dilate(backdrop_blur_padding(blur.blur_radius.0)), - ); - } - - let viewport_bounds = Bounds { - origin: point(ScaledPixels(0.0), ScaledPixels(0.0)), - size: size( - ScaledPixels::from(viewport_size.width), - ScaledPixels::from(viewport_size.height), - ), - }; - bounds = bounds.intersect(&viewport_bounds); - if bounds.is_empty() { - return None; - } - - let origin = bounds.origin.map(|component| component.floor()); - let bottom_right = bounds.bottom_right().map(|component| component.ceil()); - let bounds = Bounds::from_corners(origin, bottom_right); - Some(BackdropScratchBounds { - texture_size: size( - DevicePixels::from(bounds.size.width), - DevicePixels::from(bounds.size.height), - ), - bounds, - }) - } - - fn max_backdrop_texture_size( - scratch_bounds: BackdropScratchBounds, - viewport_size: Size, - ) -> Size { - Size { - width: DevicePixels( - (viewport_size.width.0 - scratch_bounds.bounds.origin.x.0 as i32).max(0), - ), - height: DevicePixels( - (viewport_size.height.0 - scratch_bounds.bounds.origin.y.0 as i32).max(0), - ), - } - } - - fn fit_backdrop_scratch_bounds( - mut scratch_bounds: BackdropScratchBounds, - texture_size: Size, - viewport_size: Size, - ) -> BackdropScratchBounds { - let texture_size = size( - texture_size.width.min(viewport_size.width), - texture_size.height.min(viewport_size.height), - ); - let max_origin_x = (viewport_size.width.0 - texture_size.width.0).max(0) as f32; - let max_origin_y = (viewport_size.height.0 - texture_size.height.0).max(0) as f32; - let origin = point( - ScaledPixels(scratch_bounds.bounds.origin.x.0.clamp(0.0, max_origin_x)), - ScaledPixels(scratch_bounds.bounds.origin.y.0.clamp(0.0, max_origin_y)), - ); - scratch_bounds.bounds = Bounds { - origin, - size: size( - ScaledPixels::from(texture_size.width), - ScaledPixels::from(texture_size.height), - ), - }; - scratch_bounds.texture_size = texture_size; - scratch_bounds - } - - fn prepare_backdrop_blurs( - blurs: &[BackdropBlur], - scratch_bounds: BackdropScratchBounds, - ) -> Vec { - blurs - .iter() - .cloned() - .map(|mut blur| { - blur.source_origin_x = scratch_bounds.bounds.origin.x.0; - blur.source_origin_y = scratch_bounds.bounds.origin.y.0; - blur.source_width = scratch_bounds.texture_size.width.0 as f32; - blur.source_height = scratch_bounds.texture_size.height.0 as f32; - blur - }) - .collect() - } - fn backdrop_blur_needs_source_snapshot( plan_groups: &[(usize, usize, BackdropBlurPlan)], ) -> bool { diff --git a/crates/gpui_wgpu/src/wgpu_renderer.rs b/crates/gpui_wgpu/src/wgpu_renderer.rs index a76186f..6c93a4d 100644 --- a/crates/gpui_wgpu/src/wgpu_renderer.rs +++ b/crates/gpui_wgpu/src/wgpu_renderer.rs @@ -5,8 +5,9 @@ use gpui::{ GlobalElementId, GpuSpecs, MonochromeSprite, Path, Point, PolychromeSprite, PrimitiveBatch, Quad, RetainedLayer, RetainedLayerContentRevision, ScaledPixels, Scene, Shadow, Size, SubpixelSprite, TransformationMatrix, Underline, backdrop_blur_clusters, - backdrop_blur_level_sizes_for, backdrop_blur_padding, backdrop_blur_plan_groups, - backdrop_source_bounds, can_reuse_backdrop_texture, get_gamma_correction_ratios, + backdrop_blur_level_sizes_for, backdrop_blur_plan_groups, backdrop_scratch_bounds, + backdrop_source_bounds, can_reuse_backdrop_texture, fit_backdrop_scratch_bounds, + get_gamma_correction_ratios, max_backdrop_texture_size, prepare_backdrop_blurs, }; use log::warn; #[cfg(not(target_family = "wasm"))] @@ -173,12 +174,6 @@ struct PathScratchBounds { texture_size: Size, } -#[derive(Clone, Copy)] -struct BackdropScratchBounds { - bounds: Bounds, - texture_size: Size, -} - #[derive(Clone, Copy)] #[repr(C)] struct BackdropBlurPassParams { @@ -2147,17 +2142,17 @@ impl WgpuRenderer { let mut ok = true; for blurs in blur_clusters { let Some(mut scratch_bounds) = - Self::backdrop_scratch_bounds(&blurs, target_size) + backdrop_scratch_bounds(&blurs, target_size) else { continue; }; let Some(texture_size) = self.ensure_backdrop_texture( scratch_bounds.texture_size, - Self::max_backdrop_texture_size(scratch_bounds, target_size), + max_backdrop_texture_size(scratch_bounds, target_size), ) else { continue; }; - scratch_bounds = Self::fit_backdrop_scratch_bounds( + scratch_bounds = fit_backdrop_scratch_bounds( scratch_bounds, texture_size, target_size, @@ -2165,8 +2160,7 @@ impl WgpuRenderer { let Some(backdrop_texture) = self.backdrop_texture.as_ref() else { continue; }; - let prepared_blurs = - Self::prepare_backdrop_blurs(&blurs, scratch_bounds); + let prepared_blurs = prepare_backdrop_blurs(&blurs, scratch_bounds); let plan_groups = backdrop_blur_plan_groups( &blurs, self.backdrop_blur_level_sizes.len().saturating_sub(1), @@ -2773,106 +2767,6 @@ impl WgpuRenderer { }) } - fn backdrop_scratch_bounds( - blurs: &[BackdropBlur], - viewport_size: Size, - ) -> Option { - let mut bounds = blurs - .first()? - .bounds - .dilate(backdrop_blur_padding(blurs.first()?.blur_radius.0)); - for blur in blurs.iter().skip(1) { - bounds = bounds.union( - &blur - .bounds - .dilate(backdrop_blur_padding(blur.blur_radius.0)), - ); - } - - let viewport_bounds = Bounds { - origin: Point { - x: ScaledPixels(0.0), - y: ScaledPixels(0.0), - }, - size: Size { - width: ScaledPixels::from(viewport_size.width), - height: ScaledPixels::from(viewport_size.height), - }, - }; - bounds = bounds.intersect(&viewport_bounds); - if bounds.is_empty() { - return None; - } - - let origin = bounds.origin.map(|component| component.floor()); - let bottom_right = bounds.bottom_right().map(|component| component.ceil()); - let bounds = Bounds::from_corners(origin, bottom_right); - Some(BackdropScratchBounds { - texture_size: Size { - width: DevicePixels::from(bounds.size.width), - height: DevicePixels::from(bounds.size.height), - }, - bounds, - }) - } - - fn max_backdrop_texture_size( - scratch_bounds: BackdropScratchBounds, - viewport_size: Size, - ) -> Size { - Size { - width: DevicePixels( - (viewport_size.width.0 - scratch_bounds.bounds.origin.x.0 as i32).max(0), - ), - height: DevicePixels( - (viewport_size.height.0 - scratch_bounds.bounds.origin.y.0 as i32).max(0), - ), - } - } - - fn fit_backdrop_scratch_bounds( - mut scratch_bounds: BackdropScratchBounds, - texture_size: Size, - viewport_size: Size, - ) -> BackdropScratchBounds { - let texture_size = Size { - width: texture_size.width.min(viewport_size.width), - height: texture_size.height.min(viewport_size.height), - }; - let max_origin_x = (viewport_size.width.0 - texture_size.width.0).max(0) as f32; - let max_origin_y = (viewport_size.height.0 - texture_size.height.0).max(0) as f32; - let origin = Point { - x: ScaledPixels(scratch_bounds.bounds.origin.x.0.clamp(0.0, max_origin_x)), - y: ScaledPixels(scratch_bounds.bounds.origin.y.0.clamp(0.0, max_origin_y)), - }; - scratch_bounds.bounds = Bounds { - origin, - size: Size { - width: ScaledPixels::from(texture_size.width), - height: ScaledPixels::from(texture_size.height), - }, - }; - scratch_bounds.texture_size = texture_size; - scratch_bounds - } - - fn prepare_backdrop_blurs( - blurs: &[BackdropBlur], - scratch_bounds: BackdropScratchBounds, - ) -> Vec { - blurs - .iter() - .cloned() - .map(|mut blur| { - blur.source_origin_x = scratch_bounds.bounds.origin.x.0; - blur.source_origin_y = scratch_bounds.bounds.origin.y.0; - blur.source_width = scratch_bounds.texture_size.width.0 as f32; - blur.source_height = scratch_bounds.texture_size.height.0 as f32; - blur - }) - .collect() - } - fn draw_paths_from_intermediate( &self, paths: &[Path], diff --git a/crates/gpui_windows/src/directx_renderer.rs b/crates/gpui_windows/src/directx_renderer.rs index 93c3667..f9621dd 100644 --- a/crates/gpui_windows/src/directx_renderer.rs +++ b/crates/gpui_windows/src/directx_renderer.rs @@ -128,12 +128,6 @@ struct BackdropBlurResources { upsample_srvs: Vec>, } -#[derive(Clone, Copy)] -struct BackdropScratchBounds { - bounds: Bounds, - texture_size: Size, -} - struct DirectXRenderPipelines { shadow_pipeline: PipelineState, backdrop_blur_pipeline: PipelineState, @@ -579,7 +573,7 @@ impl DirectXRenderer { ); for blurs in backdrop_blur_clusters(blurs, viewport_size) { let Some(mut scratch_bounds) = - Self::backdrop_scratch_bounds(&blurs, self.width, self.height) + backdrop_scratch_bounds(&blurs, viewport_size) else { continue; }; @@ -592,22 +586,17 @@ impl DirectXRenderer { .ensure_backdrop_resources( devices, scratch_bounds.texture_size, - Self::max_backdrop_texture_size( - scratch_bounds, - self.width, - self.height, - ), + max_backdrop_texture_size(scratch_bounds, viewport_size), )? { - scratch_bounds = Self::fit_backdrop_scratch_bounds( + scratch_bounds = fit_backdrop_scratch_bounds( scratch_bounds, texture_size, viewport_size, ); } } - let prepared_blurs = - Self::prepare_backdrop_blurs(&blurs, scratch_bounds); + let prepared_blurs = prepare_backdrop_blurs(&blurs, scratch_bounds); self.copy_render_target_to_backdrop(scratch_bounds)?; let max_backdrop_blur_levels = self.max_backdrop_blur_levels(); let mut current_plan = None; @@ -1023,98 +1012,6 @@ impl DirectXRenderer { Ok(()) } - fn backdrop_scratch_bounds( - blurs: &[BackdropBlur], - width: u32, - height: u32, - ) -> Option { - let mut bounds = blurs - .first()? - .bounds - .dilate(backdrop_blur_padding(blurs.first()?.blur_radius.0)); - for blur in blurs.iter().skip(1) { - bounds = bounds.union( - &blur - .bounds - .dilate(backdrop_blur_padding(blur.blur_radius.0)), - ); - } - - let viewport_bounds = Bounds { - origin: point(ScaledPixels(0.0), ScaledPixels(0.0)), - size: size(ScaledPixels(width as f32), ScaledPixels(height as f32)), - }; - bounds = bounds.intersect(&viewport_bounds); - if bounds.is_empty() { - return None; - } - - let origin = bounds.origin.map(|component| component.floor()); - let bottom_right = bounds.bottom_right().map(|component| component.ceil()); - let bounds = Bounds::from_corners(origin, bottom_right); - Some(BackdropScratchBounds { - texture_size: size( - DevicePixels::from(bounds.size.width), - DevicePixels::from(bounds.size.height), - ), - bounds, - }) - } - - fn max_backdrop_texture_size( - scratch_bounds: BackdropScratchBounds, - width: u32, - height: u32, - ) -> Size { - Size { - width: DevicePixels((width as i32 - scratch_bounds.bounds.origin.x.0 as i32).max(0)), - height: DevicePixels((height as i32 - scratch_bounds.bounds.origin.y.0 as i32).max(0)), - } - } - - fn fit_backdrop_scratch_bounds( - mut scratch_bounds: BackdropScratchBounds, - texture_size: Size, - viewport_size: Size, - ) -> BackdropScratchBounds { - let texture_size = size( - texture_size.width.min(viewport_size.width), - texture_size.height.min(viewport_size.height), - ); - let max_origin_x = (viewport_size.width.0 - texture_size.width.0).max(0) as f32; - let max_origin_y = (viewport_size.height.0 - texture_size.height.0).max(0) as f32; - let origin = point( - ScaledPixels(scratch_bounds.bounds.origin.x.0.clamp(0.0, max_origin_x)), - ScaledPixels(scratch_bounds.bounds.origin.y.0.clamp(0.0, max_origin_y)), - ); - scratch_bounds.bounds = Bounds { - origin, - size: size( - ScaledPixels::from(texture_size.width), - ScaledPixels::from(texture_size.height), - ), - }; - scratch_bounds.texture_size = texture_size; - scratch_bounds - } - - fn prepare_backdrop_blurs( - blurs: &[BackdropBlur], - scratch_bounds: BackdropScratchBounds, - ) -> Vec { - blurs - .iter() - .cloned() - .map(|mut blur| { - blur.source_origin_x = scratch_bounds.bounds.origin.x.0; - blur.source_origin_y = scratch_bounds.bounds.origin.y.0; - blur.source_width = scratch_bounds.texture_size.width.0 as f32; - blur.source_height = scratch_bounds.texture_size.height.0 as f32; - blur - }) - .collect() - } - fn max_backdrop_blur_levels(&self) -> usize { self.resources .as_ref() From 3d37709a26f8e8b9efb3cf481c65575684246206 Mon Sep 17 00:00:00 2001 From: Aditya Sharma Date: Mon, 27 Jul 2026 22:23:27 -0700 Subject: [PATCH 4/4] refactor(rendering): reuse first backdrop blur Bind the cluster head once so scratch-bound initialization cannot drift between duplicate lookups. --- crates/gpui/src/scene.rs | 6 +++--- 1 file changed, 3 insertions(+), 3 deletions(-) diff --git a/crates/gpui/src/scene.rs b/crates/gpui/src/scene.rs index 5269093..0a6e8ad 100644 --- a/crates/gpui/src/scene.rs +++ b/crates/gpui/src/scene.rs @@ -979,10 +979,10 @@ pub fn backdrop_scratch_bounds( blurs: &[BackdropBlur], viewport_size: Size, ) -> Option { - let mut bounds = blurs - .first()? + let first = blurs.first()?; + let mut bounds = first .bounds - .dilate(backdrop_blur_padding(blurs.first()?.blur_radius.0)); + .dilate(backdrop_blur_padding(first.blur_radius.0)); for blur in blurs.iter().skip(1) { bounds = bounds.union( &blur