diff --git a/CHANGELOG.md b/CHANGELOG.md index d619a2e475..43db29da78 100644 --- a/CHANGELOG.md +++ b/CHANGELOG.md @@ -54,6 +54,7 @@ Bottom level categories: - Bump MSRV for `d3d12`/`naga`/`wgpu-core`/`wgpu-hal`/`wgpu-types`' to 1.76. By @wumpf in [#6003](https://github.com/gfx-rs/wgpu/pull/6003) - Print requested and supported usages on `UnsupportedUsage` error. By @VladasZ in [#6007](https://github.com/gfx-rs/wgpu/pull/6007) - Fix function for checking bind compatibility to error instead of panic. By @sagudev [#6012](https://github.com/gfx-rs/wgpu/pull/6012) +- Deduplicate bind group layouts that are created from pipelines with "auto" layouts. By @teoxoy [#6049](https://github.com/gfx-rs/wgpu/pull/6049) ## 22.0.0 (2024-07-17) diff --git a/tests/tests/bind_group_layout_dedup.rs b/tests/tests/bind_group_layout_dedup.rs index 5c38779f13..591f4f9054 100644 --- a/tests/tests/bind_group_layout_dedup.rs +++ b/tests/tests/bind_group_layout_dedup.rs @@ -241,11 +241,11 @@ fn bgl_dedupe_with_dropped_user_handle(ctx: TestingContext) { } #[gpu_test] -static BIND_GROUP_LAYOUT_DEDUPLICATION_DERIVED: GpuTestConfiguration = GpuTestConfiguration::new() +static GET_DERIVED_BGL: GpuTestConfiguration = GpuTestConfiguration::new() .parameters(TestParameters::default().test_features_limits()) - .run_sync(bgl_dedupe_derived); + .run_sync(get_derived_bgl); -fn bgl_dedupe_derived(ctx: TestingContext) { +fn get_derived_bgl(ctx: TestingContext) { let buffer = ctx.device.create_buffer(&wgpu::BufferDescriptor { label: None, size: 4, @@ -314,12 +314,12 @@ fn bgl_dedupe_derived(ctx: TestingContext) { } #[gpu_test] -static SEPARATE_PROGRAMS_HAVE_INCOMPATIBLE_DERIVED_BGLS: GpuTestConfiguration = +static SEPARATE_PIPELINES_HAVE_INCOMPATIBLE_DERIVED_BGLS: GpuTestConfiguration = GpuTestConfiguration::new() .parameters(TestParameters::default().test_features_limits()) - .run_sync(separate_programs_have_incompatible_derived_bgls); + .run_sync(separate_pipelines_have_incompatible_derived_bgls); -fn separate_programs_have_incompatible_derived_bgls(ctx: TestingContext) { +fn separate_pipelines_have_incompatible_derived_bgls(ctx: TestingContext) { let buffer = ctx.device.create_buffer(&wgpu::BufferDescriptor { label: None, size: 4, @@ -448,3 +448,91 @@ fn derived_bgls_incompatible_with_regular_bgls(ctx: TestingContext) { None, ) } + +#[gpu_test] +static BIND_GROUP_LAYOUT_DEDUPLICATION_DERIVED: GpuTestConfiguration = GpuTestConfiguration::new() + .parameters(TestParameters::default().test_features_limits()) + .run_sync(bgl_dedupe_derived); + +fn bgl_dedupe_derived(ctx: TestingContext) { + let src = " + @group(0) @binding(0) var u1: vec4f; + @group(1) @binding(0) var u2: vec4f; + + @compute @workgroup_size(1, 1, 1) + fn main() { + // Just need a static use. + let _u1 = u1; + let _u2 = u2; + } + "; + let module = ctx + .device + .create_shader_module(wgpu::ShaderModuleDescriptor { + label: None, + source: wgpu::ShaderSource::Wgsl(src.into()), + }); + + let pipeline = ctx + .device + .create_compute_pipeline(&wgpu::ComputePipelineDescriptor { + label: None, + layout: None, + module: &module, + entry_point: None, + compilation_options: Default::default(), + cache: None, + }); + + let bind_group_layout_0 = pipeline.get_bind_group_layout(0); + let bind_group_layout_1 = pipeline.get_bind_group_layout(1); + + let buffer = ctx.device.create_buffer(&wgpu::BufferDescriptor { + label: None, + size: 16, + usage: wgpu::BufferUsages::UNIFORM, + mapped_at_creation: false, + }); + + let bind_group_0 = ctx.device.create_bind_group(&wgpu::BindGroupDescriptor { + label: None, + layout: &bind_group_layout_1, + entries: &[wgpu::BindGroupEntry { + binding: 0, + resource: wgpu::BindingResource::Buffer(wgpu::BufferBinding { + buffer: &buffer, + offset: 0, + size: None, + }), + }], + }); + let bind_group_1 = ctx.device.create_bind_group(&wgpu::BindGroupDescriptor { + label: None, + layout: &bind_group_layout_0, + entries: &[wgpu::BindGroupEntry { + binding: 0, + resource: wgpu::BindingResource::Buffer(wgpu::BufferBinding { + buffer: &buffer, + offset: 0, + size: None, + }), + }], + }); + + let mut encoder = ctx + .device + .create_command_encoder(&wgpu::CommandEncoderDescriptor { label: None }); + + let mut pass = encoder.begin_compute_pass(&wgpu::ComputePassDescriptor { + label: None, + timestamp_writes: None, + }); + pass.set_pipeline(&pipeline); + pass.set_bind_group(0, &bind_group_0, &[]); + pass.set_bind_group(1, &bind_group_1, &[]); + pass.dispatch_workgroups(1, 1, 1); + + drop(pass); + + ctx.queue.submit(Some(encoder.finish())); +} diff --git a/wgpu-core/src/command/bind.rs b/wgpu-core/src/command/bind.rs index 73f1d9fe17..04a992928c 100644 --- a/wgpu-core/src/command/bind.rs +++ b/wgpu-core/src/command/bind.rs @@ -176,16 +176,10 @@ mod compat { } } - #[derive(Clone, Debug, Error)] - #[error("Unknown reason")] - struct Unknown(); - Err(Error::Incompatible { expected_bgl: expected_bgl.error_ident(), assigned_bgl: assigned_bgl.error_ident(), - inner: MultiError::new(errors.drain(..)).unwrap_or_else(|| { - MultiError::new(core::iter::once(Unknown())).unwrap() - }), + inner: MultiError::new(errors.drain(..)).unwrap(), }) } } else { diff --git a/wgpu-core/src/device/bgl.rs b/wgpu-core/src/device/bgl.rs index 911ac8a435..9b7bdc0fee 100644 --- a/wgpu-core/src/device/bgl.rs +++ b/wgpu-core/src/device/bgl.rs @@ -126,4 +126,9 @@ impl EntryMap { self.sorted = false; self.inner.entry(key) } + + pub fn sort(&mut self) { + self.inner.sort_unstable_keys(); + self.sorted = true; + } } diff --git a/wgpu-core/src/device/resource.rs b/wgpu-core/src/device/resource.rs index 4a063fbf2f..9f8f48e566 100644 --- a/wgpu-core/src/device/resource.rs +++ b/wgpu-core/src/device/resource.rs @@ -32,7 +32,7 @@ use crate::{ UsageScopePool, }, validation::{self, validate_color_attachment_bytes_per_sample}, - FastHashMap, LabelHelpers as _, SubmissionIndex, + FastHashMap, LabelHelpers as _, PreHashedKey, PreHashedMap, SubmissionIndex, }; use arrayvec::ArrayVec; @@ -2592,11 +2592,29 @@ impl Device { derived_group_layouts.pop(); } + let mut unique_bind_group_layouts = PreHashedMap::default(); + let bind_group_layouts = derived_group_layouts .into_iter() - .map(|bgl_entry_map| { - self.create_bind_group_layout(&None, bgl_entry_map, bgl::Origin::Derived) - .map(Arc::new) + .map(|mut bgl_entry_map| { + bgl_entry_map.sort(); + match unique_bind_group_layouts.entry(PreHashedKey::from_key(&bgl_entry_map)) { + std::collections::hash_map::Entry::Occupied(v) => Ok(Arc::clone(v.get())), + std::collections::hash_map::Entry::Vacant(e) => { + match self.create_bind_group_layout( + &None, + bgl_entry_map, + bgl::Origin::Derived, + ) { + Ok(bgl) => { + let bgl = Arc::new(bgl); + e.insert(bgl.clone()); + Ok(bgl) + } + Err(e) => Err(e), + } + } + } }) .collect::, _>>()?; @@ -2730,11 +2748,12 @@ impl Device { if is_auto_layout { for bgl in pipeline.layout.bind_group_layouts.iter() { - bgl.exclusive_pipeline + // `bind_group_layouts` might contain duplicate entries, so we need to ignore the result. + let _ = bgl + .exclusive_pipeline .set(binding_model::ExclusivePipeline::Compute(Arc::downgrade( &pipeline, - ))) - .unwrap(); + ))); } } @@ -3355,11 +3374,12 @@ impl Device { if is_auto_layout { for bgl in pipeline.layout.bind_group_layouts.iter() { - bgl.exclusive_pipeline + // `bind_group_layouts` might contain duplicate entries, so we need to ignore the result. + let _ = bgl + .exclusive_pipeline .set(binding_model::ExclusivePipeline::Render(Arc::downgrade( &pipeline, - ))) - .unwrap(); + ))); } }