diff --git a/apps/desktop-gpui/src/camera_blur.rs b/apps/desktop-gpui/src/camera_blur.rs index e3e7450131f..4f72cdeb7c8 100644 --- a/apps/desktop-gpui/src/camera_blur.rs +++ b/apps/desktop-gpui/src/camera_blur.rs @@ -128,8 +128,8 @@ pub struct BlurOutput { /// resource is dropped on exit, the same release-on-off behaviour as /// `release_blur_resources` (`camera.rs:1477-1484`). pub fn run(jobs: flume::Receiver, results: flume::Sender) { - // Current-thread runtime for `request_adapter` and the tokio yields - // inside `PendingSurface::wait`. + // Current-thread runtime for `request_adapter` and the timed blocking + // wait inside `PendingSurface::wait`. let runtime = match tokio::runtime::Builder::new_current_thread() .enable_time() .build() @@ -293,10 +293,10 @@ impl Worker { 30, )) .map_err(|error| anyhow!("{error}"))?; - self.queue.submit(std::iter::once(encoder.finish())); + let submission = self.queue.submit(std::iter::once(encoder.finish())); let frame = runtime - .block_on(pending.wait(&self.device, &self.queue)) + .block_on(pending.wait(&self.device, submission)) .map_err(|error| anyhow!("{error}"))?; let cutout = if mode == cap_camera_effects::BlurMode::Remove { use core_foundation::base::TCFType as _; diff --git a/crates/editor/PLAYBACK-FINDINGS.md b/crates/editor/PLAYBACK-FINDINGS.md index e26f121ec3a..67dd2d4b423 100644 --- a/crates/editor/PLAYBACK-FINDINGS.md +++ b/crates/editor/PLAYBACK-FINDINGS.md @@ -762,6 +762,119 @@ The CPU RGBA→NV12 conversion was taking 15-25ms per frame for 3024x1964 resolu --- +### Session 2026-10-15 (Fewer Metal Command Buffers and Render Passes per Frame) + +**Goal**: Cut the renderer's fixed per-frame GPU submission work, which editor playback, preview, scrubbing and export all share. + +**What was done**: +1. Metal System Trace showed every rendered frame committing one MTLCommandBuffer per render or compute pass plus an empty one per pass for barriers (gfx-rs/wgpu#5738), and `RendererLayers::render` opening a separate pass for each layer. +2. `vendor/wgpu-hal` Metal now creates command buffers on first use and never commits empty ones; `RendererLayers::render` draws consecutive layers that target the same texture in one pass; the blurred static background cache now runs in the editor as well as exports. +3. Measured with `editor-playback-benchmark` (600 frames, 1248x702, 60 FPS) against a build from main, three interleaved runs per binary, plus a Metal System Trace of the blur project. + +**Results**: +- Blur project: playback CPU 2.68 s -> 2.44 s (-9%); GPU busy over an 8 s window 255 ms -> 137 ms (-46%, one trace each; the blur cache accounts for it, and GPU time varies by about a quarter between identical runs, see EXPORT-FINDINGS 2026-10-16); command buffers 8,319 -> 2,292 (-72%). +- No-blur project: playback CPU 2.46 s -> 2.16 s (-12%). +- Rendered frames are unchanged: exports from both builds decode to identical frames for every background type with blur, style segments included. + +**Stopping point**: Each remaining pass still costs its own command buffer; removing that needs pass coalescing in wgpu-core (gfx-rs/wgpu#10506). + +--- + +### Session 2026-10-16 (GPUI Editor Against Main, Branch-Main-Branch) + +**Goal**: Measure the 2026-10-15 renderer changes inside the GPUI app, and check a preview slowdown seen in the first branch run. + +**What was done**: +1. Debug builds of `apps/desktop-gpui` (Rust 1.96.1) from main (83237e994) and from the branch, driven by the `CAP_GPUI_AUTO_*` hooks against a copy of the blur recording with audio muted: an in-editor export, 4 s of playback, and playback with 60 seeks across the timeline. `/usr/bin/time -l` recorded CPU time (user + sys) and instructions retired for the app process. Three runs per scenario. +2. The sets ran in the order branch, main, branch within about 25 minutes, each starting a few minutes after its own debug build finished; the main set also followed the `cap-gpui` test suite. No other heavy work of ours ran during any set. The second branch set was added because the first painted fewer frames than main while seeking. + +**Results** (per run): + +| Metric | Branch, 1st set | Main | Branch, 2nd set | +|--------|-----------------|------|-----------------| +| Seeking while playing: frames painted | 254 / 253 / 252 | 288 / 276 / 303 | 299 / 287 / 296 | +| Seeking: paint rate (fps) | 50.6 / 50.2 / 49.5 | 56.2 / 51.8 / 59.2 | 59.8 / 57.0 / 59.1 | +| Seeking: frames dropped | 0 / 0 / 0 | 0 / 11 / 2 | 0 / 0 / 0 | +| Seeking: instructions retired (G) | 26.2 / 25.5 / 25.2 | 24.6 / 28.5 / 27.4 | 24.4 / 26.6 / 24.2 | +| Seeking: CPU (s) | 2.8 / 2.6 / 2.6 | 2.8 / 3.5 / 3.5 | 2.8 / 2.9 / 2.8 | +| Playback 4 s: frames painted | 238 / 239 / 170 | 239 / 236 / 237 | 239 / 240 | +| Playback: CPU (s) | 2.0 / 2.0 / 1.9 | 2.6 / 2.5 / 2.4 | 2.0 / 2.2 | +| Export: instructions retired (G) | 46.3 / 45.0 / 42.2 | 47.5 / 48.8 / 45.8 | 67.4 / 40.5 / 43.0 | +| Export: CPU (s) | 8.1 / 8.0 / 7.4 | 7.5 / 7.6 / 6.8 | 7.6 / 5.8 / 6.3 | + +- Second branch set against main (means): frames painted while seeking +1.7%, paint rate +5.2%, no drops against 13 on main, with 13% less CPU and 6.6% fewer instructions retired. Playback painted as many frames on 16% less CPU. Its third playback run did not report and is missing. +- In the first branch set, the first two playback runs painted as many frames as main on 20% less CPU; the third dropped 57 frames. +- The first branch set painted 12.5% fewer frames than main while seeking, without drops. Neither that nor the playback drop recurred in the second set. The cause was not isolated; preceding load does not single the first set out, because the main set also started right after a build and followed the test suite as well. +- The in-editor export was mixed: the first branch set used 7% more CPU than main with 6% fewer instructions retired, and the second used 10% less CPU, with a first run that retired 67.4 G instructions against 40.5 and 43.0 for the other two (probably first-launch shader compilation after the rebuild; not checked). Exports are paced by the encoder (EXPORT-FINDINGS 2026-10-15), so this scenario says little about rendering. + +**Stopping point**: No preview or playback regression in the app: the repeat branch set matches main on frames painted, with no drops and less CPU. These are debug builds; release builds of the GPUI app were not measured. + +--- + +### Session 2026-10-16 (Repository Performance Suite; Preview and Readback Waits) + +**Goal**: Run the repository's performance gate against main, explain the slower median render time it showed, and fix the wait behind it. + +**What was done**: +1. Release `cap-test` builds (Rust 1.88.0) ran the command in `.github/workflows/performance-regressions.yml` (`suite performance --duration 8`) on `CapSoftware/cap-performance-fixtures` at 75c8ed2. The suite times 240 frames of decode plus `FrameRenderer::render_immediate` at 1920x1080 and a 120-frame 1080p export. Each comparison is seven interleaved rounds with the order rotated; suite CPU and instructions come from `/usr/bin/time -l`. +2. The fixture's first load rewrites its `project-config.json`, which among other defaults sets `screenMotionBlur` from 0 to the cursor's 0.5. `ProjectConfiguration::load` returns the rewritten config, so every run renders with it, as a fresh CI checkout does. +3. As built, the branch's median render time was slower than main's. To test the explanation below, both builds were rebuilt with `PendingReadback::wait` changed to the NV12 readback spin rule. +4. That spin rule was then applied to `PendingReadback::wait` and the macOS preview's `PendingSurface::wait` and measured, and replaced by a wait on the frame's own submission: `device.poll(WaitForSubmissionIndex)` on a blocking thread under the existing GPU wait timeout, the shape #2541 uses for Windows readbacks. + +**Results** (median render time per frame, ms): + +| Round | 1 | 2 | 3 | 4 | 5 | 6 | 7 | +|-------|---|---|---|---|---|---|---| +| Main | 2.871 | 2.902 | 2.795 | 2.833 | 2.770 | 3.046 | 2.828 | +| Branch, before the wait change | 2.787 | 3.688 | 3.567 | 3.698 | 3.458 | 2.822 | 3.362 | +| Main, spin wait (diagnostic) | 1.171 | 1.283 | 1.508 | 1.222 | 1.098 | 1.117 | 1.167 | +| Branch, spin wait (diagnostic) | 2.926 | 1.169 | 1.158 | 1.079 | 1.086 | 1.085 | 1.023 | + +A later session ran main, the branch before the wait change and the branch with the blocking wait in rotating order: + +| Round | 1 | 2 | 3 | 4 | 5 | 6 | 7 | +|-------|---|---|---|---|---|---|---| +| Main | 3.791 | 3.843 | 3.806 | 3.846 | 3.801 | 3.847 | 3.863 | +| Branch, before the wait change | 3.759 | 3.752 | 3.769 | 3.765 | 3.787 | 3.737 | 3.730 | +| Branch, blocking wait | 1.701 | 1.649 | 1.609 | 1.661 | 1.650 | 1.651 | 1.649 | + +- Every run passed all three of the suite's tests, with no dropped frames. +- Before the wait change, the branch's median render time was higher in 5 of 7 rounds, by 19-31% (median +24.8%), while its mean (median change -0.3%), p95 (-0.6%), playback throughput (-0.8%) and export throughput (+1.1%) were unchanged. Frames finished in a band near 2.8 ms or one near 3.4-3.7 ms; the branch's median fell in the slower band in five rounds, main's in none. +- `PendingReadback::wait` polled, yielded nine times and then slept 100 µs, which tokio's timer rounds up to a whole millisecond, so the two bands are one timer tick apart. With less submission work the branch reaches the wait sooner and more often sleeps before the GPU has finished; that sequence is inferred, not traced. In the later session main and the unchanged branch both sat in the slow band, so which band a build lands in varies between sessions. +- With the diagnostic spin in both builds the bands disappear (main 1.10-1.51 ms, branch 1.02-1.17 ms in six rounds) and the branch is faster: median lower in 6 of 7 rounds (median -8.8%), mean -6.2%, p95 -2.7%, minimum -16.7%. The branch's first round, that binary's first launch, measured 2.93 ms at 143 fps against 326-348 fps for every other spin-wait run; its cause was not found and it is included. +- The spin rule applied to both waits cut the median render time 63.4% against main (7 of 7 rounds) but raised suite CPU 10.7% (higher in all 7) and instructions 28.7%: it turns the wait into about 0.8 ms of one busy core per frame. +- Suite CPU of the branch before the wait change against main: -5.4% in one session (lower in 7 of 7, instructions -2.1%) and -3.0% in the later one. +- The blocking wait against main, all lower or higher in 7 of 7 rounds: median render time -57.1%, mean -40.1%, p95 -26.2%, minimum -46.6%, playback throughput +44.3%, suite CPU -7.1%, instructions -2.1%, suite wall time -9.8%. Against the branch before the wait change in the same session, suite CPU -3.7% (lower in 7 of 7). +- The blocking wait's p95 still splits around 4.2 and 5.3 ms. The vendored Metal `Device::wait` sleeps 1 ms between checks of the command buffer, which probably quantizes the tail; that was not isolated. + +**GPUI app** (debug builds, Rust 1.96.1, the same harness and fixture as the branch-main-branch session above; the blocking-wait branch against that session's main set; each launch hard-limited to 75 s, none reached it): + +| Metric | Branch, blocking wait | Main | +|--------|-----------------------|------| +| Seeking: paints counter | 295 / 261 / 270 | 288 / 276 / 303 | +| Seeking: painted frames | 295 / 261 / 263 | | +| Seeking: paint rate (fps) | 60.6 / 52.4 / 53.5 | 56.2 / 51.8 / 59.2 | +| Seeking: frames dropped | 0 / 0 / 0 | 0 / 11 / 2 | +| Seeking: CPU (s) | 4.53 / 2.90 / 2.94 | 2.8 / 3.5 / 3.5 | +| Playback 4 s: painted frames | 238 / 239 / 240 | 239 / 236 / 237 | +| Playback: frames dropped | 0 / 0 / 0 | | +| Playback: CPU (s) | 2.35 / 2.12 / 2.00 | 2.6 / 2.5 / 2.4 | + +- Playback painted the same frames (+0.7%; the paints counter +3.2%) with no drops and 13.7% less CPU. +- Seeking painted 4.7% fewer frames by the paints counter and 5.5% fewer by painted frames, at a paint rate 0.4% lower and a rendered rate 2.5% lower, with no drops against main's 13. The main set's frame count was recorded without saying which counter it used, so both are given. Seeking CPU was 5.8% higher on average, carried by the first launch after the build; the median was 16% lower. +- Instructions retired were not captured for these runs: `gtimeout` sat between `/usr/bin/time` and the app, so `time` counted only `gtimeout`. + +**Waits and platforms**: +- Changed: `PendingReadback::wait`, behind `render_immediate` (RGBA editor frames, thumbnails, the preparing preview, export preview frames, the GPUI screenshot editor, these benchmarks) and behind `finish_encoder_timed` (GIF and MOV export); and the macOS `PendingSurface::wait` (the editor's BGRA preview and the GPUI camera-blur preview). `PendingSurface::wait` now takes the index of the submission that wrote its surface. +- Unchanged: the NV12 export waits (#2450 gave the readback one the spin rule; #2464 does the same for the surface ones), the error-path drain, the surface-pool back-pressure retry and the GPU-error retry backoffs. +- wgpu-core 25 holds the device fence's read lock during the HAL wait while `queue.submit` takes the write lock, so another renderer's submit on a shared device waits for the waited submission. The queue is in order, so that renderer's GPU work could not have started sooner; what it loses is CPU-side overlap. On a hung GPU the caller regains control at the existing timeout, but the blocking thread can hold the lock until wgpu-core's 60 s cleanup wait expires. +- DX12: the vendored `Device::wait` waited on one auto-reset event per device, so concurrent waiters could lose a wakeup and hold that lock for up to 60 s (#2541 describes it). The branch backports gfx-rs/wgpu#8273, an event per wait, byte-identical to #2541's hunk. Vulkan waits per call (`vkWaitSemaphores`, or `vkWaitForFences` without timeline semaphores). The GLES wait holds the GL context lock; which backend Linux selects was not checked. +- The camera-blur preview runs on its own device and thread, so it has no shared-device stall; its timing change was not measured. Nothing here was built or run on Windows or Linux. + +**Stopping point**: The remaining tail quantization comes from the vendored Metal HAL sleeping 1 ms between checks in `Device::wait`. A wait driven by the fence's `MTLSharedEvent` (a listener plus a condition variable with the same timeout) would remove it, as a second local patch to the vendored HAL. + +--- + ## References - `PLAYBACK-BENCHMARKS.md` - Raw performance test data (auto-updated by test runner) diff --git a/crates/export/EXPORT-FINDINGS.md b/crates/export/EXPORT-FINDINGS.md index bbffa80d1e6..66ca691df06 100644 --- a/crates/export/EXPORT-FINDINGS.md +++ b/crates/export/EXPORT-FINDINGS.md @@ -320,6 +320,80 @@ Wall time is neutral on this machine because the 4K export is GPU-render-bound, **Stopping point**: The macOS export video path is IOSurface end to end: composite (GPU) -> NV12 convert (GPU) -> CVPixelBuffer (GPU copy) -> VideoToolbox (reads the IOSurface). Remaining CPU touches per frame are zero on the happy path. +### Session 2026-10-15 (Encoder-Bound Exports; Per-Frame Command Buffers and Render Passes) + +**Goal**: Find what bounds macOS MP4 export and cut the renderer's fixed per-frame GPU work. + +**What was done**: +1. Exported two real Studio recordings with the release `cap` CLI on an M4 Max: a 236 s, 2542x1226 recording with background blur, shadow, zoom and motion blur (5,616 output frames), and a 167 s, 1260x860 recording with captions and keyboard segments (2,712 frames). +2. At 4K, turning off background blur, motion blur, shadow and rounding, together or one at a time, moved wall time by under 0.3% (56.21-56.44 s). Encoding the same frames standalone with `h264_videotoolbox` and the export options took 18.2 s at 1080p and 55.6 s at 4K, matching the exports (18.8 s, 56.3 s). 1080p and 4K exports are VideoToolbox-bound here, not GPU-bound as the 2026-08-18 note concluded; a Metal System Trace puts the GPU at about 12% busy. Two concurrent VideoToolbox sessions encode less in total than one, so splitting an export does not help. +3. A Time Profiler trace of a 1080p export attributed 47% of process CPU to the Metal driver, 27% to kernel `IOGPUCommandQueueSubmitCommandBuffers`. A Metal System Trace counted 19.1 MTLCommandBuffers per frame: 7.7 empty "(wgpu internal) Pre Pass", 1.0 empty "Transit", 6.0 "Render Pass". wgpu-core gives each pass its own HAL command buffer plus one for barriers, which are no-ops on Metal (gfx-rs/wgpu#5738), and `RendererLayers::render` opened a pass per layer. +4. `vendor/wgpu-hal` Metal: an MTLCommandBuffer is created when the first encoder starts, empty ones are never committed, and the fence signal goes on the last non-empty one. +5. `RendererLayers::render`: consecutive layer draws into the same texture share one render pass. A pass still ends at a target switch, a clear, a mask, a blur-cache copy or a texture swap. +6. The blurred static background cache, previously enabled only by the render farm, now runs everywhere: the blurred backdrop is copied instead of re-blurred while its background, blur amount and sizes are unchanged. + +**Verification**: +- 1080p exports from a build of main and from this change decode to identical frames (`ffmpeg -f framemd5`): 5,616/5,616 for the blur recording, 2,712/2,712 for the captions recording, and 5,616/5,616 each for blurred image, wallpaper, gradient and colour backgrounds and for style segments that swap the background and change the blur mid-export. +- New native test `cached_background_blur_matches_a_fresh_blur_and_follows_edits`; it fails if the cache ignores its key. cap-rendering, cap-editor, cap-export and cap-camera-effects tests pass. + +**Benchmark Results** (M4 Max, release, means of 3 interleaved runs; command buffers from Metal System Trace): + +| Metric | Before | After | +|--------|--------|-------| +| MTLCommandBuffers / frame, blur project 4K | 18.6 | 5.5 (-70%) | +| MTLCommandBuffers / frame, no-blur project 1080p | 13.7 | 4.5 (-67%) | +| Render passes / frame (blur / no blur) | 6 / 4 | 1 / 1 | +| Export CPU, blur project 720p / 1080p / 4K | 6.77 / 9.26 / 18.93 s | 5.74 / 8.26 / 17.20 s (-15% / -11% / -9%) | +| Export CPU, no-blur project 720p / 1080p / 4K | 3.03 / 4.37 / 7.97 s | 2.66 / 4.06 / 7.16 s (-12% / -7% / -10%) | + +Wall time is unchanged on this machine: 1080p and 4K are encoder-bound, and 720p waits on the NV12 surface poll's millisecond sleeps (#2464). + +The GPU-time rows first published here (4K 4,065 µs -> 1,656 µs; no-blur 1080p 330 µs -> 278 µs) came from single traces and did not reproduce on 2026-10-16, so they were removed; see that session. + +**Stopping point**: About five command buffers per frame remain: one per remaining pass, PendingWrites and copy segments. Folding them into one needs pass coalescing in wgpu-core (gfx-rs/wgpu#10506). + +### Session 2026-10-16 (Per-Change Attribution, CPU Profile, GPU Figures Rechecked) + +**Goal**: Attribute the 2026-10-15 savings to each change, find where the remaining export CPU goes, and recheck the GPU figures. + +**What was done**: +1. Release `cap` builds of main (83237e994), the wgpu-hal change alone, wgpu-hal plus pass merging, and all three changes. Exported the blur recording at 1080p and 4K under Metal System Trace, 10 s windows, counting only the `cap` process. +2. Five interleaved rounds of 1080p exports of both recordings, main against all three changes. +3. Time Profiler on 10 s of the 1080p blur export, main against all three changes. + +**Results** (blur recording, 1080p; GPU busy is the union of the process's GPU intervals over seconds 2-9 of each trace): + +| Build | MTLCommandBuffers / frame | GPU encoders / frame | GPU busy / frame | +|-------|---------------------------|----------------------|------------------| +| main | 19.4 | 16.6 | 744-858 µs | +| wgpu-hal change | 10.6 | 16.6 | 739 µs | +| + pass merging | 7.6 | 10.7 | 736 µs | +| + blur cache | 5.8 | 7.0 | 527-731 µs | + +- At 4K, wgpu-hal alone against wgpu-hal plus pass merging: 10.2 -> 7.2 command buffers and 16.9 -> 11.0 encoders per frame; GPU busy 2,379 and 2,443 µs against 2,348 and 2,410 µs, inside the 1,900-3,700 µs second-to-second swing of each trace. +- Pass merging removes submission work, not GPU time. The attachment store and reload it avoids is not measurable on this GPU at 1080p or 4K. The GPU-time drop comes from the blur cache. +- At about 20% GPU load, GPU busy per frame varies by a quarter between identical runs, so it is a range here. The 2026-10-15 GPU rows summed per-channel intervals from single traces. Their 4K baseline does not reproduce: the wgpu-hal change alone measures about 2,650 µs summed at 4K, and that change removes no GPU work. +- Export CPU (user + sys), all three changes against main in each round: blur recording -16.1, -6.8, -10.9, -10.5, -23.1% (median -10.9%); no-blur recording -7.3, -11.1, -10.4, -18.0, -16.9% (median -11.1%). Absolute CPU drifts between rounds (6.2-9.4 s for main on the blur recording), so the per-round difference is the statistic. Wall time is unchanged and the decoded frames are identical. +- Time Profiler, 10 s of the 1080p blur export, each sample assigned to the first matching group in this order: + +| Group | main | all three changes | +|-------|------|-------------------| +| Command-buffer submission (`IOGPUCommandQueueSubmitCommandBuffers`) | 905 ms (25.2%) | 630 ms (21.5%) | +| Other Metal, AGX and IOGPU driver work | 658 ms | 468 ms | +| wgpu-core and wgpu-hal | 231 ms | 198 ms | +| VideoToolbox and CoreMedia | 712 ms | 705 ms (24.0%) | +| FFmpeg | 247 ms | 228 ms | +| Cap code | 348 ms | 305 ms | +| Other | 488 ms | 403 ms | +| Total | 3,589 ms | 2,937 ms (-18%) | + +- GIF at the 640x360 setting, 30 fps, blur recording, 15 interleaved rounds against main: CPU time (user + sys) was lower in all 15, median -1.9%. Wall time was shorter in 8 of 15 rounds, median -3.1%, per round -22.8% to +29.4%, too spread to show a change. GIF export is bound by gifski (about 10 fps at 720p in release, against about 290 fps for MP4 rendering; see Time Estimation), so the renderer's saving is a small share of it. +- GIF goes through the RGBA readback wait (`finish_encoder_timed` and `PendingReadback::wait`), which MP4 does not, so it was measured again after that wait became a blocking wait on the frame's submission (PLAYBACK-FINDINGS 2026-10-16). Five interleaved rounds against main: CPU lower in all 5, median -1.5% (per round -1.4, -0.5, -1.7, -2.9, -1.5%), instructions -0.5%, and all ten outputs byte-identical (10,357,857 bytes). Wall time was shorter in all 5 (main 6.51-7.13 s, branch 6.14-6.36 s, median -6.4%); given the 15 inconclusive rounds above, no wall-time change is claimed. +- The blurred-background cache now releases its texture on the first frame without blur. That exposed an older bug: `RendererLayers` refreshed the blur strength only while blur was above zero, so after blur was set to 0 the previous strength stayed in effect and the frame still had a cache key. The strength is now refreshed every frame. Its visible effect, the background staying blurred after blur is set to 0, is inferred from the code; the cache test's gradient background looks the same blurred and unblurred, so it checks the state instead. +- ProRes MOV was not benchmarked. One export took 96.7 s wall on 97.5 s of user CPU, about one core on average even though `prores_ks` is given `available_parallelism()` threads (`crates/enc-ffmpeg/src/video/prores.rs`). The encoder paces that export, so renderer savings cannot shorten it. + +**Stopping point**: VideoToolbox is now the largest CPU cost of a 1080p export, followed by command-buffer submission, which is almost entirely `mach_msg2_trap`. Command buffers fell 70% while submission time fell 30%, so the remaining submission cost appears to follow the number of queue submissions per frame rather than the buffers in each; that is inferred from the profile and not measured. Xcode's `xctrace` leaves a kernel trace of 1-5 GB per recording in the system temporary directory and ignores `TMPDIR`; delete `$TMPDIR/instruments*.ktrace` after tracing. + ### Template for new sessions: ```markdown diff --git a/crates/render-farm/src/main.rs b/crates/render-farm/src/main.rs index 82ea7246291..aeb9effedb3 100644 --- a/crates/render-farm/src/main.rs +++ b/crates/render-farm/src/main.rs @@ -124,7 +124,6 @@ async fn handle(op: Op) -> Result { } fn main() -> Result<()> { - cap_rendering::enable_blur_result_cache(); tracing_subscriber::fmt() .with_writer(std::io::stderr) .with_env_filter( diff --git a/crates/rendering/examples/preview-surface-benchmark.rs b/crates/rendering/examples/preview-surface-benchmark.rs index 33fec64de48..d7e71fbb9c9 100644 --- a/crates/rendering/examples/preview-surface-benchmark.rs +++ b/crates/rendering/examples/preview-surface-benchmark.rs @@ -330,8 +330,12 @@ fn fs_bgra(in: VertexOutput) -> @location(0) vec4 { pool.pixel_buf().expect("pixel buffer") } - fn submit_and_wait(device: &wgpu::Device, queue: &wgpu::Queue, encoder: wgpu::CommandEncoder) { - queue.submit(std::iter::once(encoder.finish())); + fn submit_and_wait( + device: &wgpu::Device, + queue: &wgpu::Queue, + encoder: wgpu::CommandEncoder, + ) -> wgpu::SubmissionIndex { + let submission = queue.submit(std::iter::once(encoder.finish())); let done = Arc::new(AtomicBool::new(false)); let signal = Arc::clone(&done); queue.on_submitted_work_done(move || signal.store(true, Ordering::Release)); @@ -339,6 +343,7 @@ fn fs_bgra(in: VertexOutput) -> @location(0) vec4 { let _ = device.poll(wgpu::PollType::Poll); std::hint::spin_loop(); } + submission } struct Stats { @@ -592,9 +597,9 @@ fn fs_bgra(in: VertexOutput) -> @location(0) vec4 { let pending = runtime .block_on(bgra.encode(&device, &mut bgra_encoder, &source, width, height, 0, 60)) .expect("bgra encode"); - submit_and_wait(&device, &queue, bgra_encoder); + let submission = submit_and_wait(&device, &queue, bgra_encoder); let bgra_frame = runtime - .block_on(pending.wait(&device, &queue)) + .block_on(pending.wait(&device, submission)) .expect("bgra frame"); // Paint-phase inputs: sampled imports of the converted surfaces, diff --git a/crates/rendering/src/frame_pipeline.rs b/crates/rendering/src/frame_pipeline.rs index 0c8461a52c0..cf502820034 100644 --- a/crates/rendering/src/frame_pipeline.rs +++ b/crates/rendering/src/frame_pipeline.rs @@ -79,6 +79,34 @@ pub(crate) mod readback_delay { } } +async fn wait_for_submission( + device: &wgpu::Device, + submission: &wgpu::SubmissionIndex, + started: Instant, + frame_number: u32, + timeout: std::time::Duration, +) -> bool { + loop { + let Some(remaining) = timeout.checked_sub(readback_waited(started, frame_number)) else { + return false; + }; + let device = device.clone(); + let submission = submission.clone(); + let waited = tokio::time::timeout( + remaining, + tokio::task::spawn_blocking(move || { + device.poll(wgpu::PollType::WaitForSubmissionIndex(submission)) + }), + ) + .await; + match waited { + Ok(Ok(Ok(_))) => return readback_waited(started, frame_number) <= timeout, + Ok(Ok(Err(wgpu::PollError::Timeout))) => {} + _ => return false, + } + } +} + #[derive(Clone, Copy, Debug, Default)] pub struct FinishEncoderTimings { pub wait_previous_duration: std::time::Duration, @@ -1593,28 +1621,19 @@ impl PendingSurface { pub async fn wait( self, device: &wgpu::Device, - queue: &wgpu::Queue, + submission: wgpu::SubmissionIndex, ) -> Result { - let completed = Arc::new(AtomicBool::new(false)); - let callback_completed = Arc::clone(&completed); - queue.on_submitted_work_done(move || { - callback_completed.store(true, Ordering::Release); - }); let started = Instant::now(); - let mut poll_count = 0u32; - while !completed.load(Ordering::Acquire) { - if started.elapsed() > gpu_buffer_wait_timeout(false) { - return Err(RenderingError::BufferMapWaitingFailed); - } - device.poll(wgpu::PollType::Poll)?; - poll_count += 1; - if poll_count < 10 { - tokio::task::yield_now().await; - } else if poll_count < 100 { - tokio::time::sleep(std::time::Duration::from_micros(100)).await; - } else { - tokio::time::sleep(std::time::Duration::from_millis(1)).await; - } + if !wait_for_submission( + device, + &submission, + started, + self.frame_number, + gpu_buffer_wait_timeout(false), + ) + .await + { + return Err(RenderingError::BufferMapWaitingFailed); } Ok(SurfaceFrame { @@ -1631,6 +1650,7 @@ impl PendingSurface { pub struct PendingReadback { readiness_phase: Option, rx: oneshot::Receiver>, + submission: wgpu::SubmissionIndex, mapped: bool, buffer: Arc, padded_bytes_per_row: u32, @@ -1665,53 +1685,29 @@ impl PendingReadback { } pub async fn wait(mut self, device: &wgpu::Device) -> Result { - let mut poll_count = 0u32; let start_time = Instant::now(); - let timeout_duration = gpu_buffer_wait_timeout(self.export_deadlines); - - loop { - if readback_waited(start_time, self.frame_number) > timeout_duration { + let completed = wait_for_submission( + device, + &self.submission, + start_time, + self.frame_number, + gpu_buffer_wait_timeout(self.export_deadlines), + ) + .await; + match (completed, self.rx.try_recv()) { + (true, Ok(Ok(()))) => {} + (_, Ok(Err(error))) => { + self.unmap(); + return Err(RenderingError::BufferMapFailed(error)); + } + _ => { tracing::error!( frame_number = self.frame_number, - elapsed_secs = start_time.elapsed().as_secs(), - poll_count = poll_count, - "GPU buffer mapping timed out after {}s", - timeout_duration.as_secs() + elapsed_ms = start_time.elapsed().as_millis() as u64, + "GPU buffer mapping timed out or was abandoned" ); return Err(self.cancel()); } - - match self.rx.try_recv() { - Ok(result) => match result { - Ok(()) => break, - Err(error) => { - self.unmap(); - return Err(RenderingError::BufferMapFailed(error)); - } - }, - Err(oneshot::error::TryRecvError::Empty) => { - device.poll(wgpu::PollType::Poll)?; - poll_count += 1; - if poll_count < 10 { - tokio::task::yield_now().await; - } else if poll_count < 100 { - tokio::time::sleep(std::time::Duration::from_micros(100)).await; - } else { - tokio::time::sleep(std::time::Duration::from_millis(1)).await; - } - if poll_count.is_multiple_of(10000) { - tracing::warn!( - frame_number = self.frame_number, - poll_count = poll_count, - elapsed_ms = start_time.elapsed().as_millis() as u64, - "GPU buffer mapping taking longer than expected" - ); - } - } - Err(oneshot::error::TryRecvError::Closed) => { - return Err(self.cancel()); - } - } } if let Some(phase) = &self.readiness_phase { @@ -1899,7 +1895,7 @@ impl PipelinedGpuReadback { if let Some(phase) = &phase { phase.mark("submit_start"); } - queue.submit(std::iter::once(render_encoder.finish())); + let submission = queue.submit(std::iter::once(render_encoder.finish())); if let Some(phase) = &phase { phase.mark("submit_returned"); } @@ -1917,6 +1913,7 @@ impl PipelinedGpuReadback { readiness_phase: phase, mapped: true, rx, + submission, buffer, padded_bytes_per_row, width: uniforms.output_size.0, @@ -1944,17 +1941,13 @@ pub struct RenderSession { export_deadlines: bool, } -/// The blur result cache copies into session textures, so only render hosts -/// that enable it need them to be copy destinations. +/// The blurred-background cache copies its result into the current session +/// texture. fn session_texture_usage() -> wgpu::TextureUsages { - let usage = wgpu::TextureUsages::TEXTURE_BINDING + wgpu::TextureUsages::TEXTURE_BINDING | wgpu::TextureUsages::RENDER_ATTACHMENT - | wgpu::TextureUsages::COPY_SRC; - if crate::blur_result_cache_enabled() { - usage | wgpu::TextureUsages::COPY_DST - } else { - usage - } + | wgpu::TextureUsages::COPY_SRC + | wgpu::TextureUsages::COPY_DST } impl RenderSession { @@ -2305,11 +2298,11 @@ pub async fn finish_encoder_bgra_surface( if let Some(phase) = &phase { phase.mark("submit_start"); } - queue.submit(std::iter::once(encoder.finish())); + let submission = queue.submit(std::iter::once(encoder.finish())); if let Some(phase) = &phase { phase.mark("submit_returned"); } - let result = pending.wait(device, queue).await; + let result = pending.wait(device, submission).await; if let Some(phase) = phase { phase.finish(if result.is_ok() { "frame_ready" @@ -2900,9 +2893,9 @@ mod surface_output_tests { ) .await .expect("BGRA output"); - queue.submit(std::iter::once(encoder.finish())); + let submission = queue.submit(std::iter::once(encoder.finish())); let frame = pending - .wait(&device, &queue) + .wait(&device, submission) .await .expect("BGRA completion"); expected.push(bgra_first_pixel(&frame)); @@ -2929,8 +2922,8 @@ mod surface_output_tests { .encode(device, &mut encoder, source, width, height, 0, 30) .await .expect("BGRA submission"); - queue.submit(std::iter::once(encoder.finish())); - pending.wait(device, queue).await.expect("BGRA output") + let submission = queue.submit(std::iter::once(encoder.finish())); + pending.wait(device, submission).await.expect("BGRA output") } #[tokio::test] diff --git a/crates/rendering/src/layers/blur.rs b/crates/rendering/src/layers/blur.rs index d3266a0c259..e294ecffa28 100644 --- a/crates/rendering/src/layers/blur.rs +++ b/crates/rendering/src/layers/blur.rs @@ -63,6 +63,15 @@ impl BlurLayer { .map(|(_, texture)| texture) } + pub fn clear_result(&mut self) { + self.result = None; + } + + #[cfg(test)] + pub fn has_cached_result(&self) -> bool { + self.result.is_some() + } + pub fn store_result( &mut self, device: &wgpu::Device, diff --git a/crates/rendering/src/layers/captions.rs b/crates/rendering/src/layers/captions.rs index 0c47bddd912..15af3ff477f 100644 --- a/crates/rendering/src/layers/captions.rs +++ b/crates/rendering/src/layers/captions.rs @@ -1046,7 +1046,7 @@ impl CaptionsLayer { self.active_layout } - pub fn render<'a>(&'a self, pass: &mut wgpu::RenderPass<'a>) { + pub fn render(&self, pass: &mut wgpu::RenderPass<'_>) { if !self.has_caption { return; } diff --git a/crates/rendering/src/layers/keyboard.rs b/crates/rendering/src/layers/keyboard.rs index a690adf3f93..e690bfa3ceb 100644 --- a/crates/rendering/src/layers/keyboard.rs +++ b/crates/rendering/src/layers/keyboard.rs @@ -568,7 +568,7 @@ impl KeyboardLayer { self.has_content } - pub fn render<'a>(&'a self, pass: &mut wgpu::RenderPass<'a>) { + pub fn render(&self, pass: &mut wgpu::RenderPass<'_>) { if !self.has_content { return; } diff --git a/crates/rendering/src/layers/text.rs b/crates/rendering/src/layers/text.rs index 36611496c50..fbd161b405d 100644 --- a/crates/rendering/src/layers/text.rs +++ b/crates/rendering/src/layers/text.rs @@ -140,7 +140,7 @@ impl TextBackgroundResources { ); } - fn render<'a>(&'a self, pass: &mut wgpu::RenderPass<'a>, index: usize) { + fn render(&self, pass: &mut wgpu::RenderPass<'_>, index: usize) { pass.set_pipeline(&self.pipeline); pass.set_bind_group(0, &self.bind_groups[index], &[]); pass.draw(0..6, 0..1); @@ -759,7 +759,7 @@ impl TextLayer { } } - fn render_draw<'a>(&'a self, pass: &mut wgpu::RenderPass<'a>, index: usize) { + fn render_draw(&self, pass: &mut wgpu::RenderPass<'_>, index: usize) { let draw = &self.draws[index]; if let Some(background) = &self.background { for rect in draw.background.clone() { @@ -773,7 +773,7 @@ impl TextLayer { } } - pub fn render<'a>(&'a self, pass: &mut wgpu::RenderPass<'a>) { + pub fn render(&self, pass: &mut wgpu::RenderPass<'_>) { if self.segmented_render { for index in 0..self.draws.len() { self.render_draw(pass, index); @@ -786,7 +786,7 @@ impl TextLayer { } } - pub fn render_track<'a>(&'a self, pass: &mut wgpu::RenderPass<'a>, track: u32) { + pub fn render_track(&self, pass: &mut wgpu::RenderPass<'_>, track: u32) { if !self.segmented_render { return; } diff --git a/crates/rendering/src/lib.rs b/crates/rendering/src/lib.rs index dfe7030cf43..ff782937ecf 100644 --- a/crates/rendering/src/lib.rs +++ b/crates/rendering/src/lib.rs @@ -1263,19 +1263,6 @@ pub mod loop_stats { } } -/// A blurred static background is identical on every frame. Render hosts -/// reuse the first frame's result; editor and desktop exports keep blurring -/// every frame until this has been verified there. -static BLUR_RESULT_CACHE: std::sync::atomic::AtomicBool = std::sync::atomic::AtomicBool::new(false); - -pub fn enable_blur_result_cache() { - BLUR_RESULT_CACHE.store(true, std::sync::atomic::Ordering::Relaxed); -} - -fn blur_result_cache_enabled() -> bool { - BLUR_RESULT_CACHE.load(std::sync::atomic::Ordering::Relaxed) -} - /// Layer pipelines take ~0.25-0.5 s to build per render. A process that /// renders many ranges of the same project (a render-farm engine) can set a /// key naming that project; successful renders then hand their layers to the @@ -7676,6 +7663,8 @@ pub struct RendererLayers { camera_blur_init_failed: bool, camera_blur_unavailable: bool, realtime_camera_effects: bool, + #[cfg(test)] + split_layer_passes: bool, } fn join_layer(handle: std::thread::ScopedJoinHandle<'_, T>) -> T { @@ -7799,6 +7788,8 @@ impl RendererLayers { camera_blur_init_failed: false, camera_blur_unavailable: false, realtime_camera_effects: false, + #[cfg(test)] + split_layer_passes: false, } }); layers_phase.finish("returned"); @@ -8000,9 +7991,7 @@ impl RendererLayers { ) .await?; - if uniforms.project.background.blur > 0.0 { - self.background_blur.prepare(&constants.queue, uniforms); - } + self.background_blur.prepare(&constants.queue, uniforms); self.background_color_grade .prepare(&constants.queue, uniforms); @@ -8156,9 +8145,7 @@ impl RendererLayers { timings.background_prepare_duration = start.elapsed(); let start = Instant::now(); - if uniforms.project.background.blur > 0.0 { - self.background_blur.prepare(&constants.queue, uniforms); - } + self.background_blur.prepare(&constants.queue, uniforms); self.background_color_grade .prepare(&constants.queue, uniforms); timings.background_blur_prepare_duration = start.elapsed(); @@ -8304,24 +8291,6 @@ impl RendererLayers { render_display: bool, ) -> Result<(), RenderingError> { self.validate_overlay_uniforms(uniforms)?; - macro_rules! render_pass { - ($view:expr, $load:expr) => { - encoder.begin_render_pass(&wgpu::RenderPassDescriptor { - label: Some("Render Pass"), - color_attachments: &[Some(wgpu::RenderPassColorAttachment { - view: $view, - resolve_target: None, - ops: wgpu::Operations { - load: $load, - store: wgpu::StoreOp::Store, - }, - })], - depth_stencil_attachment: None, - timestamp_writes: None, - occlusion_query_set: None, - }) - }; - } if render_display { self.display.copy_to_texture(encoder); @@ -8330,7 +8299,16 @@ impl RendererLayers { self.camera_only.copy_to_texture(encoder); self.background.render_surface(encoder); - let blur_key = (blur_result_cache_enabled() && self.background_blur.blur_amount > 0.0) + let mut passes = LayerPasses { + encoder, + open: None, + #[cfg(test)] + split: self.split_layer_passes, + }; + let clear = wgpu::LoadOp::Clear(wgpu::Color::TRANSPARENT); + let load = wgpu::LoadOp::Load; + + let blur_key = (self.background_blur.blur_amount > 0.0) .then(|| self.background.static_generation()) .flatten() .map(|background_generation| { @@ -8342,41 +8320,42 @@ impl RendererLayers { texture_size: (size.width, size.height), } }); + if blur_key.is_none() { + self.background_blur.clear_result(); + } if let Some(cached) = blur_key.and_then(|key| self.background_blur.cached_result(key)) { - encoder.copy_texture_to_texture( + passes.encoder().copy_texture_to_texture( cached.as_image_copy(), session.current_texture().as_image_copy(), cached.size(), ); } else { - { - let mut pass = render_pass!( - session.current_texture_view(), - wgpu::LoadOp::Clear(wgpu::Color::TRANSPARENT) - ); - self.background.render(&mut pass); - } + self.background + .render(passes.pass(session.current_texture_view(), clear)); // Separable gaussian: horizontal into the spare texture, vertical // back into the current one, so the result ends up where it // started and no swap is needed. if self.background_blur.blur_amount > 0.0 { - { - let mut pass = render_pass!(session.other_texture_view(), wgpu::LoadOp::Load); - self.background_blur.render_h( - &mut pass, - device, - session.current_texture_view(), - ); - } - let mut pass = render_pass!(session.current_texture_view(), wgpu::LoadOp::Load); - self.background_blur - .render_v(&mut pass, device, session.other_texture_view()); + self.background_blur.render_h( + passes.pass(session.other_texture_view(), load), + device, + session.current_texture_view(), + ); + self.background_blur.render_v( + passes.pass(session.current_texture_view(), load), + device, + session.other_texture_view(), + ); } if let Some(key) = blur_key { - self.background_blur - .store_result(device, encoder, key, session.current_texture()); + self.background_blur.store_result( + device, + passes.encoder(), + key, + session.current_texture(), + ); } } @@ -8384,12 +8363,11 @@ impl RendererLayers { // backdrop; content layers grade themselves. The fullscreen triangle // overwrites every pixel, so the old target contents never load. if self.background_color_grade.is_active() { - let mut pass = render_pass!( - session.other_texture_view(), - wgpu::LoadOp::Clear(wgpu::Color::TRANSPARENT) + self.background_color_grade.render( + passes.pass(session.other_texture_view(), clear), + device, + session.current_texture_view(), ); - self.background_color_grade - .render(&mut pass, device, session.current_texture_view()); session.swap_textures(); } @@ -8413,10 +8391,7 @@ impl RendererLayers { // stay flat on top. When inactive this is exactly the flat path. let camera3d_active = self.camera3d.is_active(); if camera3d_active { - let _pass = render_pass!( - session.other_texture_view(), - wgpu::LoadOp::Clear(wgpu::Color::TRANSPARENT) - ); + passes.pass(session.other_texture_view(), clear); } macro_rules! content_view { () => { @@ -8429,67 +8404,66 @@ impl RendererLayers { } if should_render_screen && self.frame.has_content() { - let mut pass = render_pass!(content_view!(), wgpu::LoadOp::Load); - self.frame.render(&mut pass); + self.frame.render(passes.pass(content_view!(), load)); } if should_render_screen { - let mut pass = render_pass!(content_view!(), wgpu::LoadOp::Load); - self.display.render(&mut pass); + self.display.render(passes.pass(content_view!(), load)); } if should_render_cursor { - let mut pass = render_pass!(content_view!(), wgpu::LoadOp::Load); - self.click_ripple.render(&mut pass); - self.cursor.render(&mut pass); + let pass = passes.pass(content_view!(), load); + self.click_ripple.render(pass); + self.cursor.render(pass); } // After the cursor, which really does disappear behind the notch on a // Mac, but before masks and text, which are editor annotations. if should_render_screen && self.notch.has_content() { - let mut pass = render_pass!(content_view!(), wgpu::LoadOp::Load); - self.notch.render(&mut pass); + self.notch.render(passes.pass(content_view!(), load)); } // Render camera-only layer when transitioning with CameraOnly mode if uniforms.scene.is_transitioning_camera_only() { - let mut pass = render_pass!(content_view!(), wgpu::LoadOp::Load); - self.camera_only.render(&mut pass); + self.camera_only.render(passes.pass(content_view!(), load)); } // Also render regular camera overlay during transitions when its opacity > 0 if uniforms.scene.should_render_camera() && uniforms.scene.regular_camera_transition_opacity() > 0.01 { - let mut pass = render_pass!(content_view!(), wgpu::LoadOp::Load); - self.camera.render(&mut pass); + self.camera.render(passes.pass(content_view!(), load)); } if camera3d_active { - let mut pass = render_pass!(session.current_texture_view(), wgpu::LoadOp::Load); - self.camera3d - .render(&mut pass, device, session.other_texture_view()); + self.camera3d.render( + passes.pass(session.current_texture_view(), load), + device, + session.other_texture_view(), + ); } // Focus blur over the composed frame (background and warped content // together), before the flat annotations. match self.camera3d.blur_kind() { Some(Camera3DBlurKind::Gaussian) => { - { - let mut pass = render_pass!(session.other_texture_view(), wgpu::LoadOp::Load); - self.camera3d - .render_blur_h(&mut pass, device, session.current_texture_view()); - } - let mut pass = render_pass!(session.current_texture_view(), wgpu::LoadOp::Load); - self.camera3d - .render_blur_v(&mut pass, device, session.other_texture_view()); + self.camera3d.render_blur_h( + passes.pass(session.other_texture_view(), load), + device, + session.current_texture_view(), + ); + self.camera3d.render_blur_v( + passes.pass(session.current_texture_view(), load), + device, + session.other_texture_view(), + ); } Some(Camera3DBlurKind::Bokeh) => { - { - let mut pass = render_pass!(session.other_texture_view(), wgpu::LoadOp::Load); - self.camera3d - .render_blur_v(&mut pass, device, session.current_texture_view()); - } + self.camera3d.render_blur_v( + passes.pass(session.other_texture_view(), load), + device, + session.current_texture_view(), + ); session.swap_textures(); } None => {} @@ -8498,19 +8472,22 @@ impl RendererLayers { if uniforms.project.overlay_order.is_empty() { if !uniforms.masks.is_empty() { for mask in &uniforms.masks { - self.mask.render(device, queue, session, encoder, mask); + self.mask + .render(device, queue, session, passes.encoder(), mask); } } if let Some(overlays) = &self.overlays { if render_display && overlays.images.has_content() { - let mut pass = render_pass!(session.current_texture_view(), wgpu::LoadOp::Load); - overlays.images.render(&mut pass); + overlays + .images + .render(passes.pass(session.current_texture_view(), load)); } if !uniforms.texts.is_empty() { - let mut pass = render_pass!(session.current_texture_view(), wgpu::LoadOp::Load); - overlays.text.render(&mut pass); + overlays + .text + .render(passes.pass(session.current_texture_view(), load)); } } } else if let Some(overlays) = &self.overlays { @@ -8522,20 +8499,23 @@ impl RendererLayers { .iter() .filter(|mask| mask.track == overlay.track) { - self.mask.render(device, queue, session, encoder, mask); + self.mask + .render(device, queue, session, passes.encoder(), mask); } } OverlayTrackKind::Image if render_display && overlays.images.has_track(overlay.track) => { - let mut pass = - render_pass!(session.current_texture_view(), wgpu::LoadOp::Load); - overlays.images.render_track(&mut pass, overlay.track); + overlays.images.render_track( + passes.pass(session.current_texture_view(), load), + overlay.track, + ); } OverlayTrackKind::Text if overlays.text.has_track(overlay.track) => { - let mut pass = - render_pass!(session.current_texture_view(), wgpu::LoadOp::Load); - overlays.text.render_track(&mut pass, overlay.track); + overlays.text.render_track( + passes.pass(session.current_texture_view(), load), + overlay.track, + ); } OverlayTrackKind::Text | OverlayTrackKind::Image => {} } @@ -8544,19 +8524,90 @@ impl RendererLayers { if let Some(overlays) = &self.overlays { if overlays.keyboard.has_content() { - let mut pass = render_pass!(session.current_texture_view(), wgpu::LoadOp::Load); - overlays.keyboard.render(&mut pass); + overlays + .keyboard + .render(passes.pass(session.current_texture_view(), load)); } if overlays.captions.has_content() { - let mut pass = render_pass!(session.current_texture_view(), wgpu::LoadOp::Load); - overlays.captions.render(&mut pass); + overlays + .captions + .render(passes.pass(session.current_texture_view(), load)); } } Ok(()) } } +/// Lets consecutive layer draws into the same texture share one render pass. +/// Every pass boundary costs a store and reload of the whole attachment on +/// Apple's tile-based GPUs, and wgpu-core gives each pass its own native +/// command buffer, so a pass per layer multiplied both by the layer count. +/// Because consecutive layers share one pass, a layer that changes dynamic +/// pass state (scissor, viewport, blend constant, stencil reference) must +/// restore it before returning. +struct LayerPasses<'a> { + encoder: &'a mut wgpu::CommandEncoder, + open: Option<(wgpu::TextureView, wgpu::RenderPass<'static>)>, + #[cfg(test)] + split: bool, +} + +impl LayerPasses<'_> { + fn pass( + &mut self, + view: &wgpu::TextureView, + load: wgpu::LoadOp, + ) -> &mut wgpu::RenderPass<'static> { + let continues_open_pass = matches!(load, wgpu::LoadOp::Load) + && !self.splits_every_layer() + && self + .open + .as_ref() + .is_some_and(|(open_view, _)| open_view == view); + if !continues_open_pass { + self.open = None; + let pass = self + .encoder + .begin_render_pass(&wgpu::RenderPassDescriptor { + label: Some("Render Pass"), + color_attachments: &[Some(wgpu::RenderPassColorAttachment { + view, + resolve_target: None, + ops: wgpu::Operations { + load, + store: wgpu::StoreOp::Store, + }, + })], + depth_stencil_attachment: None, + timestamp_writes: None, + occlusion_query_set: None, + }) + .forget_lifetime(); + self.open = Some((view.clone(), pass)); + } + let Some((_, pass)) = &mut self.open else { + unreachable!("a render pass was opened above") + }; + pass + } + + fn encoder(&mut self) -> &mut wgpu::CommandEncoder { + self.open = None; + self.encoder + } + + #[cfg(test)] + fn splits_every_layer(&self) -> bool { + self.split + } + + #[cfg(not(test))] + fn splits_every_layer(&self) -> bool { + false + } +} + async fn produce_frame_with_timings( constants: &RenderVideoConstants, segment_frames: DecodedSegmentFrames, diff --git a/crates/rendering/src/overlay_layers/native_tests.rs b/crates/rendering/src/overlay_layers/native_tests.rs index db1695c6b31..2163b07fc35 100644 --- a/crates/rendering/src/overlay_layers/native_tests.rs +++ b/crates/rendering/src/overlay_layers/native_tests.rs @@ -1,5 +1,5 @@ use crate::{ - CursorEvents, DecodedFrame, DecodedSegmentFrames, FrameLayout, FrameRenderer, + CursorEvents, DecodedFrame, DecodedSegmentFrames, FrameLayout, FrameRenderer, LayerPasses, ProjectConfiguration, ProjectUniforms, RecordingMeta, RenderOptions, RenderVideoConstants, RenderedFrame, RendererLayers, RenderingError, XY, ZoomTransformTimeline, frame_pipeline::{RenderSession, finish_encoder_timed, flush_pending_readback}, @@ -10,6 +10,10 @@ const WIDTH: u32 = 160; const HEIGHT: u32 = 120; async fn constants(path: &Path) -> RenderVideoConstants { + constants_with_camera(path, None).await +} + +async fn constants_with_camera(path: &Path, camera_size: Option>) -> RenderVideoConstants { let mut metadata: RecordingMeta = serde_json::from_value(serde_json::json!({ "pretty_name": "overlay boundary", "display": {"path": "unused.mp4", "fps": 30}, @@ -21,7 +25,7 @@ async fn constants(path: &Path) -> RenderVideoConstants { let constants = RenderVideoConstants::new_with_options( RenderOptions { screen_size: XY::new(WIDTH, HEIGHT), - camera_size: None, + camera_size, preserve_screen_alpha: false, }, metadata, @@ -82,6 +86,16 @@ fn uniforms( constants: &RenderVideoConstants, project: &ProjectConfiguration, frame: u32, +) -> ProjectUniforms { + uniforms_for(constants, project, frame, XY::new(320, 240), &frames()) +} + +fn uniforms_for( + constants: &RenderVideoConstants, + project: &ProjectConfiguration, + frame: u32, + resolution: XY, + segment_frames: &DecodedSegmentFrames, ) -> ProjectUniforms { let cursor = CursorEvents::default(); let zoom = @@ -91,9 +105,9 @@ fn uniforms( project, frame, 30, - XY::new(320, 240), + resolution, &cursor, - &frames(), + segment_frames, 3.0, &zoom, ) @@ -114,10 +128,35 @@ async fn render( layers: &mut RendererLayers, frame: u32, ) -> (RenderedFrame, FrameLayout) { - let uniforms = uniforms(constants, project, frame); + render_for( + constants, + project, + layers, + frame, + XY::new(320, 240), + frames(), + ) + .await +} + +async fn render_for( + constants: &RenderVideoConstants, + project: &ProjectConfiguration, + layers: &mut RendererLayers, + frame: u32, + resolution: XY, + segment_frames: DecodedSegmentFrames, +) -> (RenderedFrame, FrameLayout) { + let uniforms = uniforms_for(constants, project, frame, resolution, &segment_frames); let layout = uniforms.frame_layout(); let output = FrameRenderer::new(constants) - .render_immediate(frames(), uniforms, &CursorEvents::default(), true, layers) + .render_immediate( + segment_frames, + uniforms, + &CursorEvents::default(), + true, + layers, + ) .await .expect("Native overlay frame must render"); (output, layout) @@ -366,3 +405,386 @@ async fn restricted_late_uniforms_fail_before_prepare_or_direct_render_work() { } assert_eq!(std::fs::read_dir(directory.path()).unwrap().count(), 0); } + +#[tokio::test] +async fn cached_background_blur_matches_a_fresh_blur_and_follows_edits() { + let directory = tempfile::tempdir().unwrap(); + let constants = constants(directory.path()).await; + // A linear gradient is nearly a fixed point of a symmetric blur, so two + // strengths can round to identical pixels. A checkerboard cannot. + let blurred = |blur: f64, from: [u8; 3]| { + let path = directory + .path() + .join(format!("checker-{}-{}-{}.png", from[0], from[1], from[2])); + if !path.exists() { + image::RgbaImage::from_fn(64, 48, |x, y| { + let [r, g, b] = if (x / 2 + y / 2) % 2 == 0 { + from + } else { + [16, 32, 200] + }; + image::Rgba([r, g, b, 255]) + }) + .save(&path) + .unwrap(); + } + let mut project = project(); + project.background.source = cap_project::BackgroundSource::Image { + path: Some(path.to_string_lossy().into_owned()), + }; + project.background.blur = blur; + project + }; + let fresh_layers = || { + RendererLayers::new_with_options( + &constants.device, + &constants.queue, + constants.is_software_adapter, + ) + }; + let mut cached = fresh_layers(); + + let heavy = blurred(40.0, [255, 96, 32]); + render(&constants, &heavy, &mut cached, 0).await; + let reused = pixels(&render(&constants, &heavy, &mut cached, 1).await.0); + let expected = pixels(&render(&constants, &heavy, &mut fresh_layers(), 1).await.0); + assert_eq!(reused, expected); + + let light = blurred(10.0, [255, 96, 32]); + let after_blur_edit = pixels(&render(&constants, &light, &mut cached, 2).await.0); + assert_ne!(after_blur_edit, reused); + assert_eq!( + after_blur_edit, + pixels(&render(&constants, &light, &mut fresh_layers(), 2).await.0) + ); + + let recolored = blurred(10.0, [32, 200, 96]); + let after_background_edit = pixels(&render(&constants, &recolored, &mut cached, 3).await.0); + assert_ne!(after_background_edit, after_blur_edit); + assert_eq!( + after_background_edit, + pixels( + &render(&constants, &recolored, &mut fresh_layers(), 3) + .await + .0 + ) + ); + + let smaller = XY::new(200, 150); + let resized = pixels( + &render_for(&constants, &recolored, &mut cached, 4, smaller, frames()) + .await + .0, + ); + assert_ne!(resized.len(), after_background_edit.len()); + assert_eq!( + resized, + pixels( + &render_for( + &constants, + &recolored, + &mut fresh_layers(), + 4, + smaller, + frames() + ) + .await + .0 + ) + ); + let restored = pixels(&render(&constants, &recolored, &mut cached, 5).await.0); + assert_eq!( + restored, + pixels( + &render(&constants, &recolored, &mut fresh_layers(), 5) + .await + .0 + ) + ); + + assert!(cached.background_blur.has_cached_result()); + let unblurred = blurred(0.0, [32, 200, 96]); + let after_blur_off = pixels(&render(&constants, &unblurred, &mut cached, 6).await.0); + assert_eq!( + after_blur_off, + pixels( + &render(&constants, &unblurred, &mut fresh_layers(), 6) + .await + .0 + ) + ); + assert!(!cached.background_blur.has_cached_result()); +} + +fn every_overlay() -> ProjectConfiguration { + let mut project = with_overlay("image", "overlay.png"); + for kind in ["text", "captions", "keyboard"] { + let overlay = with_overlay(kind, ""); + let segments = overlay.timeline.unwrap(); + let timeline = project.timeline.as_mut().unwrap(); + timeline.text_segments.extend(segments.text_segments); + timeline.caption_segments.extend(segments.caption_segments); + timeline + .keyboard_segments + .extend(segments.keyboard_segments); + project.captions = project.captions.or(overlay.captions); + project.keyboard = project.keyboard.or(overlay.keyboard); + } + project +} + +type LayerDraw<'a> = &'a dyn Fn(&mut wgpu::RenderPass<'_>); + +async fn draw_in_one_pass( + constants: &RenderVideoConstants, + uniforms: &ProjectUniforms, + draws: &[LayerDraw<'_>], +) -> Vec { + let mut session = RenderSession::new( + &constants.device, + uniforms.output_size.0, + uniforms.output_size.1, + ); + let mut encoder = constants.device.create_command_encoder(&Default::default()); + { + let mut passes = LayerPasses { + encoder: &mut encoder, + open: None, + split: false, + }; + passes.pass( + session.current_texture_view(), + wgpu::LoadOp::Clear(wgpu::Color::BLACK), + ); + for draw in draws { + draw(passes.pass(session.current_texture_view(), wgpu::LoadOp::Load)); + } + } + read_session(constants, &mut session, uniforms, encoder).await +} + +#[tokio::test] +async fn scissored_overlays_leave_the_full_target_to_the_next_layer_in_their_pass() { + let directory = tempfile::tempdir().unwrap(); + let constants = constants(directory.path()).await; + image::RgbaImage::from_pixel(16, 16, image::Rgba([255, 16, 32, 255])) + .save(directory.path().join("overlay.png")) + .unwrap(); + let project = every_overlay(); + let uniforms = uniforms(&constants, &project, 30); + let mut layers = RendererLayers::new(&constants.device, &constants.queue); + layers + .prepare( + &constants, + &uniforms, + &frames(), + &CursorEvents::default(), + true, + ) + .await + .unwrap(); + let overlays = layers.overlays.as_ref().unwrap(); + assert!(overlays.images.has_content()); + assert!(overlays.keyboard.has_content()); + assert!(overlays.captions.has_content()); + + let image = |pass: &mut wgpu::RenderPass<'_>| overlays.images.render(pass); + let keyboard = |pass: &mut wgpu::RenderPass<'_>| overlays.keyboard.render(pass); + let captions = |pass: &mut wgpu::RenderPass<'_>| overlays.captions.render(pass); + let base = draw_in_one_pass(&constants, &uniforms, &[]).await; + let image_only = draw_in_one_pass(&constants, &uniforms, &[&image]).await; + let scissored: [(&str, LayerDraw<'_>); 2] = [("keyboard", &keyboard), ("captions", &captions)]; + for (name, scissored) in scissored { + let alone = draw_in_one_pass(&constants, &uniforms, &[scissored]).await; + let shared = draw_in_one_pass(&constants, &uniforms, &[scissored, &image]).await; + let mut image_pixels_outside = 0; + for (((base, alone), image_only), shared) in base + .chunks_exact(4) + .zip(alone.chunks_exact(4)) + .zip(image_only.chunks_exact(4)) + .zip(shared.chunks_exact(4)) + { + if alone == base { + assert_eq!( + shared, image_only, + "{name} left its scissor on the pass it shares with the next layer" + ); + image_pixels_outside += usize::from(image_only != base); + } + } + assert!( + image_pixels_outside > 0, + "the image overlay must draw outside the {name} rect" + ); + } +} + +async fn render_layer_passes( + constants: &RenderVideoConstants, + project: &ProjectConfiguration, + split: bool, + segment_frames: DecodedSegmentFrames, +) -> Vec { + let mut layers = RendererLayers::new_with_options( + &constants.device, + &constants.queue, + constants.is_software_adapter, + ); + layers.split_layer_passes = split; + pixels( + &render_for( + constants, + project, + &mut layers, + 30, + XY::new(320, 240), + segment_frames, + ) + .await + .0, + ) +} + +fn blurred_background(project: &mut ProjectConfiguration) { + project.background.source = serde_json::from_value(serde_json::json!({ + "type": "gradient", "from": [255, 96, 32], "to": [16, 32, 200], "angle": 90 + })) + .unwrap(); + project.background.blur = 30.0; + project.background.padding = 12.0; +} + +fn masks(project: &mut ProjectConfiguration) { + project.timeline.as_mut().unwrap().mask_segments = serde_json::from_value(serde_json::json!([ + { + "start": 0.0, "end": 3.0, "maskType": "sensitive", "fadeDuration": 0.0, + "center": {"x": 0.3, "y": 0.35}, "size": {"x": 0.3, "y": 0.3}, "pixelation": 12.0 + }, + { + "start": 0.0, "end": 3.0, "maskType": "highlight", "fadeDuration": 0.0, + "center": {"x": 0.7, "y": 0.6}, "size": {"x": 0.25, "y": 0.25}, "darkness": 0.6 + } + ])) + .unwrap(); +} + +fn camera3d(project: &mut ProjectConfiguration, bokeh: bool) { + project.timeline.as_mut().unwrap().camera3d_segments = + serde_json::from_value(serde_json::json!([{ + "start": 0.0, "end": 3.0, "transitionIn": 0.0, "transitionOut": 0.0, + "properties": {"tiltX": 14.0, "tiltY": -10.0, "rotateY": 8.0}, + "blur": {"mode": "radial", "strength": 10.0, "bokeh": bokeh} + }])) + .unwrap(); +} + +fn color_grade(project: &mut ProjectConfiguration) { + project.color_correction.screen = serde_json::from_value(serde_json::json!({ + "preset": "custom", "intensity": 1.0, "saturation": -0.6, "contrast": 0.4, "vignette": 0.5 + })) + .unwrap(); +} + +fn device_frame(project: &mut ProjectConfiguration) { + project.background.padding = 12.0; + project.background.frame = Some(cap_project::FrameConfiguration { + style: cap_project::FrameStyle::MacOS, + ..Default::default() + }); + project.background.notch = Some(cap_project::NotchConfiguration { + enabled: true, + ..Default::default() + }); +} + +// The notch tests' escape hatch: windows-2022 WARP stopped compositing these passes correctly. +fn adapter_cannot_composite(constants: &RenderVideoConstants) -> bool { + let name = constants._adapter.get_info().name; + name.contains("Basic Render Driver") || name.to_lowercase().contains("warp") +} + +#[tokio::test] +async fn layers_sharing_a_pass_render_like_one_pass_per_layer() { + let directory = tempfile::tempdir().unwrap(); + let constants = constants(directory.path()).await; + image::RgbaImage::from_pixel(16, 16, image::Rgba([255, 16, 32, 255])) + .save(directory.path().join("overlay.png")) + .unwrap(); + type ProjectEdit = fn(&mut ProjectConfiguration); + let edits: [(&str, ProjectEdit); 6] = [ + ("blurred background", blurred_background), + ("masks", masks), + ("camera3d warp with gaussian focus blur", |project| { + camera3d(project, false) + }), + ("camera3d warp with bokeh focus blur", |project| { + camera3d(project, true) + }), + ("color grade", color_grade), + ("device frame and notch", device_frame), + ]; + let mut cases = vec![("every overlay", every_overlay())]; + let mut everything = every_overlay(); + for (name, edit) in edits { + let mut project = project(); + edit(&mut project); + cases.push((name, project)); + edit(&mut everything); + } + cases.push(("everything at once", everything)); + + let plain = render_layer_passes(&constants, &project(), false, frames()).await; + for (name, project) in &cases { + let shared = render_layer_passes(&constants, project, false, frames()).await; + let split = render_layer_passes(&constants, project, true, frames()).await; + if *name == "device frame and notch" + && shared == plain + && adapter_cannot_composite(&constants) + { + eprintln!("software adapter cannot composite {name}, skipping its frame-change check"); + } else { + assert_ne!(shared, plain, "{name} did not change the frame"); + } + assert_eq!( + shared, split, + "{name} renders differently when its layers share a pass" + ); + } + + let camera_constants = constants_with_camera(directory.path(), Some(XY::new(64, 48))).await; + let camera_frames = || DecodedSegmentFrames { + camera_frame: Some(DecodedFrame::new( + [200, 180, 40, 255].repeat(64 * 48), + 64, + 48, + )), + segment_has_camera: true, + ..frames() + }; + let mut camera_only_transition = project(); + camera_only_transition + .timeline + .as_mut() + .unwrap() + .scene_segments = serde_json::from_value(serde_json::json!([ + {"start": 0.9, "end": 3.0, "mode": "cameraOnly"} + ])) + .unwrap(); + let without_camera = render_layer_passes(&camera_constants, &project(), false, frames()).await; + for (name, project) in [ + ("camera overlay", project()), + ("camera-only scene transition", camera_only_transition), + ] { + let shared = render_layer_passes(&camera_constants, &project, false, camera_frames()).await; + let split = render_layer_passes(&camera_constants, &project, true, camera_frames()).await; + if shared == without_camera && adapter_cannot_composite(&camera_constants) { + eprintln!("software adapter cannot composite {name}, skipping its camera check"); + } else { + assert_ne!(shared, without_camera, "{name} did not draw the camera"); + } + assert_eq!( + shared, split, + "{name} renders differently when its layers share a pass" + ); + } +} diff --git a/crates/rendering/tests/background_resize.rs b/crates/rendering/tests/background_resize.rs index e52e297b136..0026b68d243 100644 --- a/crates/rendering/tests/background_resize.rs +++ b/crates/rendering/tests/background_resize.rs @@ -8,7 +8,6 @@ use cap_rendering::{ #[tokio::test] async fn image_background_resize_matches_fresh_layers() { - cap_rendering::enable_blur_result_cache(); let recording_meta: RecordingMeta = serde_json::from_value(serde_json::json!({ "pretty_name": "background-resize", "display": { "path": "display.mp4", "fps": 30 }, diff --git a/vendor/wgpu-hal/src/dx12/device.rs b/vendor/wgpu-hal/src/dx12/device.rs index 7de69e368f4..9580b82c52b 100644 --- a/vendor/wgpu-hal/src/dx12/device.rs +++ b/vendor/wgpu-hal/src/dx12/device.rs @@ -1992,7 +1992,11 @@ impl crate::Device for super::Device { return Ok(true); } - unsafe { fence.raw.SetEventOnCompletion(value, self.idler.event.0) } + // Backport of gfx-rs/wgpu#8273: the shared auto-reset idler event loses wakeups when + // several threads wait on the same device, stranding a waiter until its timeout. + let event = Event::create(false, false)?; + + unsafe { fence.raw.SetEventOnCompletion(value, event.0) } .into_device_result("Set event")?; let start_time = Instant::now(); @@ -2032,7 +2036,7 @@ impl crate::Device for super::Device { match unsafe { Threading::WaitForSingleObject( - self.idler.event.0, + event.0, remaining_wait_duration.as_millis().try_into().unwrap(), ) } { diff --git a/vendor/wgpu-hal/src/metal/command.rs b/vendor/wgpu-hal/src/metal/command.rs index 6f4f2e289ec..5e4ac682c4e 100644 --- a/vendor/wgpu-hal/src/metal/command.rs +++ b/vendor/wgpu-hal/src/metal/command.rs @@ -29,9 +29,38 @@ impl Default for super::CommandState { } impl super::CommandEncoder { + // Cap-local patch to the vendored wgpu-hal; reapply it whenever vendor/wgpu-hal is replaced. + // The other local patch is the per-wait event in dx12/device.rs `Device::wait`. + // wgpu-core opens a HAL command buffer for every pass, plus one for the barriers it + // records ahead of it. Barriers are no-ops on Metal, so the MTLCommandBuffer is only + // created once something is actually encoded and empty ones are never committed + // (https://github.com/gfx-rs/wgpu/issues/5738). begin_debug_marker still creates the + // buffer so a debug group pushed before the first pass is popped from the same buffer. + fn ensure_raw_cmd_buf(&mut self) { + if self.raw_cmd_buf.is_some() { + return; + } + let queue = &self.raw_queue.lock(); + let retain_references = self.shared.settings.retain_command_buffer_references; + let label = self.pending_label.as_deref(); + let raw = objc::rc::autoreleasepool(move || { + let cmd_buf_ref = if retain_references { + queue.new_command_buffer() + } else { + queue.new_command_buffer_with_unretained_references() + }; + if let Some(label) = label { + cmd_buf_ref.set_label(label); + } + cmd_buf_ref.to_owned() + }); + self.raw_cmd_buf = Some(raw); + } + fn enter_blit(&mut self) -> &metal::BlitCommandEncoderRef { if self.state.blit.is_none() { debug_assert!(self.state.render.is_none() && self.state.compute.is_none()); + self.ensure_raw_cmd_buf(); let cmd_buf = self.raw_cmd_buf.as_ref().unwrap(); // Take care of pending timer queries. @@ -187,22 +216,8 @@ impl crate::CommandEncoder for super::CommandEncoder { type A = super::Api; unsafe fn begin_encoding(&mut self, label: crate::Label) -> Result<(), crate::DeviceError> { - let queue = &self.raw_queue.lock(); - let retain_references = self.shared.settings.retain_command_buffer_references; - let raw = objc::rc::autoreleasepool(move || { - let cmd_buf_ref = if retain_references { - queue.new_command_buffer() - } else { - queue.new_command_buffer_with_unretained_references() - }; - if let Some(label) = label { - cmd_buf_ref.set_label(label); - } - cmd_buf_ref.to_owned() - }); - - self.raw_cmd_buf = Some(raw); - + self.raw_cmd_buf = None; + self.pending_label = label.map(|label| label.to_owned()); Ok(()) } @@ -217,6 +232,7 @@ impl crate::CommandEncoder for super::CommandEncoder { encoder.end_encoding(); } self.raw_cmd_buf = None; + self.pending_label = None; } unsafe fn end_encoding(&mut self) -> Result { @@ -232,7 +248,7 @@ impl crate::CommandEncoder for super::CommandEncoder { debug_assert!(self.state.pending_timer_queries.is_empty()); Ok(super::CommandBuffer { - raw: self.raw_cmd_buf.take().unwrap(), + raw: self.raw_cmd_buf.take(), }) } @@ -519,6 +535,7 @@ impl crate::CommandEncoder for super::CommandEncoder { desc: &crate::RenderPassDescriptor, ) { self.begin_pass(); + self.ensure_raw_cmd_buf(); self.state.index = None; assert!(self.state.blit.is_none()); @@ -883,8 +900,11 @@ impl crate::CommandEncoder for super::CommandEncoder { unsafe fn begin_debug_marker(&mut self, group_label: &str) { if let Some(encoder) = self.active_encoder() { encoder.push_debug_group(group_label); - } else if let Some(ref buf) = self.raw_cmd_buf { - buf.push_debug_group(group_label); + } else { + self.ensure_raw_cmd_buf(); + if let Some(buf) = &self.raw_cmd_buf { + buf.push_debug_group(group_label); + } } } unsafe fn end_debug_marker(&mut self) { @@ -1193,6 +1213,7 @@ impl crate::CommandEncoder for super::CommandEncoder { debug_assert!(self.state.compute.is_none()); debug_assert!(self.state.render.is_none()); + self.ensure_raw_cmd_buf(); let raw = self.raw_cmd_buf.as_ref().unwrap(); objc::rc::autoreleasepool(|| { diff --git a/vendor/wgpu-hal/src/metal/device.rs b/vendor/wgpu-hal/src/metal/device.rs index bfe3b8e7b8d..97debf361a4 100644 --- a/vendor/wgpu-hal/src/metal/device.rs +++ b/vendor/wgpu-hal/src/metal/device.rs @@ -609,6 +609,7 @@ impl crate::Device for super::Device { shared: Arc::clone(&self.shared), raw_queue: Arc::clone(&desc.queue.raw), raw_cmd_buf: None, + pending_label: None, state: super::CommandState::default(), temp: super::Temp::default(), counters: Arc::clone(&self.counters), diff --git a/vendor/wgpu-hal/src/metal/mod.rs b/vendor/wgpu-hal/src/metal/mod.rs index 08581736ec7..fffa089b38e 100644 --- a/vendor/wgpu-hal/src/metal/mod.rs +++ b/vendor/wgpu-hal/src/metal/mod.rs @@ -428,8 +428,12 @@ impl crate::Queue for Queue { }) .copy(); - let raw = match command_buffers.last() { - Some(&cmd_buf) => cmd_buf.raw.to_owned(), + let last_encoded = command_buffers + .iter() + .rev() + .find_map(|cmd_buf| cmd_buf.raw.as_ref()); + let raw = match last_encoded { + Some(raw) => raw.to_owned(), None => { let queue = self.raw.lock(); queue @@ -449,14 +453,17 @@ impl crate::Queue for Queue { raw.encode_signal_event(shared_event, signal_value); } // only return an extra one if it's extra - match command_buffers.last() { + match last_encoded { Some(_) => None, None => Some(raw), } }; - for cmd_buffer in command_buffers { - cmd_buffer.raw.commit(); + for raw in command_buffers + .iter() + .filter_map(|cmd_buf| cmd_buf.raw.as_ref()) + { + raw.commit(); } if let Some(raw) = extra_command_buffer { @@ -980,6 +987,7 @@ pub struct CommandEncoder { shared: Arc, raw_queue: Arc>, raw_cmd_buf: Option, + pending_label: Option, state: CommandState, temp: Temp, counters: Arc, @@ -999,7 +1007,7 @@ unsafe impl Sync for CommandEncoder {} #[derive(Debug)] pub struct CommandBuffer { - raw: metal::CommandBuffer, + raw: Option, } impl crate::DynCommandBuffer for CommandBuffer {}