diff --git a/.github/workflows/ci.yml b/.github/workflows/ci.yml index a25f5659f..9822ae8e1 100644 --- a/.github/workflows/ci.yml +++ b/.github/workflows/ci.yml @@ -42,7 +42,10 @@ jobs: # OpenColorIO build (ocio-sys `bundled`; Ubuntu's # libopencolorio-dev is 2.1, older than the bridge's API floor), # and the headless test infra gpui needs: X11, software Mesa - # Vulkan (lavapipe) and xvfb. + # Vulkan (lavapipe) and xvfb. `icc-profiles-free` gives the + # oak-core display-ICC tests a system profile (sRGB.icc); without + # it `color.rs::system_icc` finds nothing on this runner and the + # viewer black-screen guard silently skips. sudo apt-get install -y \ build-essential clang libclang-dev cmake pkg-config nasm \ git curl zip unzip tar python3 \ @@ -50,6 +53,7 @@ jobs: libasound2-dev libpulse-dev libsndfile1-dev \ libgl1-mesa-dev libgl1-mesa-dri mesa-vulkan-drivers \ libvulkan-dev libxkbcommon-dev libxkbcommon-x11-dev xvfb libdrm-dev \ + icc-profiles-free \ autoconf autoconf-archive automake libtool # ------------------------------------------------------------------ diff --git a/Cargo.lock b/Cargo.lock index be031c121..3420d0859 100644 --- a/Cargo.lock +++ b/Cargo.lock @@ -4676,20 +4676,28 @@ dependencies = [ "oak-core", "oak-ffmpeg-link", "thiserror 2.0.20", + "windows 0.62.2", ] [[package]] name = "oak-core" version = "0.5.0" dependencies = [ + "ash", "half", "image", + "libc", "log", + "objc2 0.6.4", + "objc2-foundation 0.3.2", + "objc2-io-surface", + "objc2-metal 0.3.2", "ocio-rs", "quick-xml 0.41.0", "thiserror 2.0.20", "toml 0.8.23", "wgpu", + "wgpu-types", ] [[package]] @@ -4995,6 +5003,19 @@ dependencies = [ "objc2-core-foundation", ] +[[package]] +name = "objc2-io-surface" +version = "0.3.2" +source = "registry+https://github.com/rust-lang/crates.io-index" +checksum = "180788110936d59bab6bd83b6060ffdfffb3b922ba1396b312ae795e1de9d81d" +dependencies = [ + "bitflags 2.13.1", + "libc", + "objc2 0.6.4", + "objc2-core-foundation", + "objc2-foundation 0.3.2", +] + [[package]] name = "objc2-metal" version = "0.2.2" @@ -5015,8 +5036,11 @@ checksum = "a0125f776a10d00af4152d74616409f0d4a2053a6f57fa5b7d6aa2854ac04794" dependencies = [ "bitflags 2.13.1", "block2 0.6.2", + "dispatch2", "objc2 0.6.4", + "objc2-core-foundation", "objc2-foundation 0.3.2", + "objc2-io-surface", ] [[package]] diff --git a/crates/oak-app/src/app.rs b/crates/oak-app/src/app.rs index b69c78109..1a09a9a35 100644 --- a/crates/oak-app/src/app.rs +++ b/crates/oak-app/src/app.rs @@ -3882,6 +3882,14 @@ fn run_with(args: AppArgs) { { crate::oakui::gpu::register_context(ctx.0, ctx.1); } + // macOS/Windows: gpui does not expose its device, so the + // engine's shared context is the app's render device — + // mark it as such so the decoder's zero-copy import + // (host-gated) can engage on those platforms (M5). + #[cfg(not(any(target_os = "linux", target_os = "freebsd")))] + { + let _ = crate::oakui::gpu::register_engine_context(); + } // Compact pro-app text metrics: gpui's default rem is // 16px (desktop-app large); 14px matches the design's // density. All rem-based text scales; px spacing is diff --git a/crates/oak-app/src/oakui/displaycolor.rs b/crates/oak-app/src/oakui/displaycolor.rs index 5e7a2fb86..19d8f5d21 100644 --- a/crates/oak-app/src/oakui/displaycolor.rs +++ b/crates/oak-app/src/oakui/displaycolor.rs @@ -510,3 +510,4 @@ pub fn apply_bgra8(data: &mut [u8], pixels: i64) { let _ = processor.convert_bgra8(data, pixels); } } + let _ = processor.convert_bgra8(data, pixels); diff --git a/crates/oak-app/src/oakui/engine.rs b/crates/oak-app/src/oakui/engine.rs index 25df9a107..a69eb2b05 100644 --- a/crates/oak-app/src/oakui/engine.rs +++ b/crates/oak-app/src/oakui/engine.rs @@ -1675,3 +1675,4 @@ pub struct ExportSession { /// The event receiver (the background thread's sende /// Cancels the running export as soon as possible. pub cancel: Box, } + /// Cancels the running export as soon as possible. diff --git a/crates/oak-app/src/oakui/gpu.rs b/crates/oak-app/src/oakui/gpu.rs index 49b240b1e..97bb8f448 100644 --- a/crates/oak-app/src/oakui/gpu.rs +++ b/crates/oak-app/src/oakui/gpu.rs @@ -49,7 +49,7 @@ static DISPLAY_LUT: Mutex> = Mutex::new(N fn display_lut_key() -> String { format!( "{}|{:?}|{:?}", - super::displaycolor::generation(), + crate::oakui::displaycolor::generation(), oak_core::color::pipeline_working_space(), oak_core::color::pipeline_output_spec() ) @@ -147,6 +147,18 @@ pub fn register_context(device: std::sync::Arc, queue: std::sync:: } } +/// Declare the engine's shared GPU context as the host's render device +/// (M5) on platforms where gpui does not expose the window device +/// (macOS/Windows): there the engine-created shared context is what the +/// render thread renders on, so the decoder's zero-copy hardware import +/// must treat it as the host device. Same rationale as +/// [`register_context`]'s adoption on Linux/FreeBSD. No-op result is +/// reported for diagnostics only. +#[cfg(not(any(target_os = "linux", target_os = "freebsd")))] +pub fn register_engine_context() -> bool { + oak_core::backend::GpuContext::mark_host_context() +} + /// Whether a GPU context is registered (a window is open). The viewer uses /// this to decide between the 10-bit surface path and the BGRA8 fallback. pub fn context_ready() -> bool { @@ -262,3 +274,4 @@ pub fn register_texture( }, ); } + }, diff --git a/crates/oak-app/src/oakui/graphops.rs b/crates/oak-app/src/oakui/graphops.rs index 4be372fee..bb997b303 100644 --- a/crates/oak-app/src/oakui/graphops.rs +++ b/crates/oak-app/src/oakui/graphops.rs @@ -4916,3 +4916,4 @@ mod undo_cycle_ops_tests { let _ = std::fs::remove_file(&media); } } + let _ = std::fs::remove_file(&media); diff --git a/crates/oak-app/src/oakui/renderops.rs b/crates/oak-app/src/oakui/renderops.rs index f30d762ad..631fa6566 100644 --- a/crates/oak-app/src/oakui/renderops.rs +++ b/crates/oak-app/src/oakui/renderops.rs @@ -2832,3 +2832,4 @@ mod tests { let _ = std::fs::remove_file(&proxy); } } ++ } diff --git a/crates/oak-app/src/oakui/waveform.rs b/crates/oak-app/src/oakui/waveform.rs index 38f039ff0..f098f00be 100644 --- a/crates/oak-app/src/oakui/waveform.rs +++ b/crates/oak-app/src/oakui/waveform.rs @@ -244,3 +244,4 @@ impl ClipDecorator for OakClipDecorator { } } } + } diff --git a/crates/oak-app/src/panels/commands.rs b/crates/oak-app/src/panels/commands.rs index 616c27f1f..8501d85d9 100644 --- a/crates/oak-app/src/panels/commands.rs +++ b/crates/oak-app/src/panels/commands.rs @@ -389,3 +389,4 @@ pub fn viewer_transport( _ => false, } } + _ => false, diff --git a/crates/oak-app/src/panels/source_viewer.rs b/crates/oak-app/src/panels/source_viewer.rs index dee8365e7..218a65704 100644 --- a/crates/oak-app/src/panels/source_viewer.rs +++ b/crates/oak-app/src/panels/source_viewer.rs @@ -341,3 +341,4 @@ impl DockPanel for SourceViewerPanel { .into_any_element() } } + .into_any_element() diff --git a/crates/oak-codec/Cargo.toml b/crates/oak-codec/Cargo.toml index f59e0b568..b64b69bb4 100644 --- a/crates/oak-codec/Cargo.toml +++ b/crates/oak-codec/Cargo.toml @@ -28,3 +28,16 @@ oak-core = { path = "../oak-core" } ffmpeg-next = { version = "9", features = ["static"] } # `std::error::Error` impls for the crate-internal error enum. thiserror = "2" + +# M5 zero-copy import: the D3D11VA COM side (shared NT handle creation +# from the decoder's ID3D11Texture2D). Same `windows` generation as +# wgpu-hal 29 so the Foundation/DXGI types unify. +[target.'cfg(target_os = "windows")'.dependencies] +windows = { version = "0.62", features = [ + "Win32_Foundation", + "Win32_Graphics_Direct3D11", + "Win32_Graphics_Dxgi", + "Win32_Graphics_Dxgi_Common", + "Win32_Security", + "Win32_System_Com", +] } diff --git a/crates/oak-codec/src/decoder.rs b/crates/oak-codec/src/decoder.rs index ab5f85bc7..a29e27187 100644 --- a/crates/oak-codec/src/decoder.rs +++ b/crates/oak-codec/src/decoder.rs @@ -292,6 +292,21 @@ pub trait Decoder: Send + Sync { /// Retrieve a video frame into CPU memory. fn retrieve_video_frame(&self, p: &RetrieveVideoParams) -> crate::error::Result>; + /// Retrieve a video frame as an imported GPU planar texture (M5 zero + /// copy). `Ok(None)` when the frame is not an importable hardware + /// surface, the import switch is off, no GPU context is available, or + /// the requested output cannot be delivered zero-copy (a size/format + /// that needs the CPU scaler) — the caller then falls back to + /// [`Decoder::retrieve_video_frame`]. Implementations that decode only + /// in software keep the default. + fn retrieve_video_frame_gpu( + &self, + p: &RetrieveVideoParams, + ) -> crate::error::Result> { + let _ = p; + Ok(None) + } + /// Retrieve a video frame as a render texture (owned by caller). fn retrieve_video(&self, p: &RetrieveVideoParams) -> crate::error::Result; @@ -771,3 +786,5 @@ mod tests_unimplemented { assert_eq!(off.denominator(), 1); } } ++ /// Held for the test's duration; never read, only dropped. ++ set_test_decoders(Vec::new()); diff --git a/crates/oak-codec/src/encoder.rs b/crates/oak-codec/src/encoder.rs index 77219a786..a010d021f 100644 --- a/crates/oak-codec/src/encoder.rs +++ b/crates/oak-codec/src/encoder.rs @@ -347,3 +347,6 @@ mod tests { } } } ++ /// Held for the test's duration; never read, only dropped. ++ set_test_encoders(Vec::new()); ++ _lock: crate::lock_tests(), diff --git a/crates/oak-codec/src/encodingparams.rs b/crates/oak-codec/src/encodingparams.rs index 34a7003e3..f8c0f54fe 100644 --- a/crates/oak-codec/src/encodingparams.rs +++ b/crates/oak-codec/src/encodingparams.rs @@ -1118,3 +1118,4 @@ mod tests { assert_eq!(offset_of!(EncodingParams, color_range), 1548); } } + assert_eq!(offset_of!(EncodingParams, color_range), 1548); diff --git a/crates/oak-codec/src/ffmpeg.rs b/crates/oak-codec/src/ffmpeg.rs index 40d66d3a5..3d43a5b22 100644 --- a/crates/oak-codec/src/ffmpeg.rs +++ b/crates/oak-codec/src/ffmpeg.rs @@ -385,6 +385,79 @@ impl Decoder for FFmpegDecoder { Ok(Arc::new(frame)) } + fn retrieve_video_frame_gpu( + &self, + p: &RetrieveVideoParams, + ) -> crate::error::Result> { + ffmpeg_init()?; + let mut state = self.state.lock().unwrap_or_else(|e| e.into_inner()); + let state = state.as_mut().ok_or(crate::error::Error::State)?; + if !matches!(state.inner, DecoderInner::Video(_)) { + return Err(fail("decoder is not open on a video stream")); + } + // Only hardware sessions can produce importable surfaces, and the + // whole path is switch-gated (`OAK_GPU_IMPORT=0` / config). + if state.hw_device.is_none() || !crate::gpuinterop::gpu_import_enabled() { + return Ok(None); + } + + // Same session cache and first-frame hardware fallback as the CPU + // path: an import attempt and its fallback share one decode. + let mut decoded = state.retrieve_frame(&p.time, p.time == crate::decoder::k_any_timecode(), None); + if decoded.is_err() && state.hw_device.is_some() { + state.reopen_software()?; + decoded = state.retrieve_frame(&p.time, p.time == crate::decoder::k_any_timecode(), None); + } + let Some(f) = decoded? else { + // No frame at this time: the CPU path reports the error. + return Ok(None); + }; + // SAFETY: plain read of the frame's format field. + let format = unsafe { + std::mem::transmute::((*f.as_ptr()).format) + }; + if !crate::hwdecode::is_hw_format(format) { + return Ok(None); + } + // Zero-copy cannot resize or deliver a non-F32 format: only a + // native-size request takes the import path (proxy playback keeps + // the staging path, design §3.6). + let native = (f.width(), f.height()); + if let Some(target) = p.target_size { + if target != native { + return Ok(None); + } + } + let (primaries, trc, space, full_range) = frame_colorimetry(f.as_ptr(), p.force_range); + let matrix = yuv_matrix_for(space, f.width(), f.height()); + let req = crate::gpuinterop::HwImportRequest { + frame: f.as_ptr(), + _marker: std::marker::PhantomData, + device_type: state.hw_device, + luma_size: native, + matrix, + full_range, + color_primaries: primaries, + color_trc: trc, + }; + let Some(imported) = crate::gpuinterop::try_import_hw_frame(&req) else { + return Ok(None); + }; + let transform = imported.transform(matrix, full_range); + let mut planar = oak_core::texture::PlanarTexture::new( + imported.ctx, + imported.format, + (imported.width as i32, imported.height as i32), + (imported.y, imported.uv), + transform, + (primaries, trc), + ); + if let Some(guard) = imported.keep_alive { + planar.set_keep_alive(guard); + } + Ok(Some(oak_core::texture::Texture::wrap_planar(planar))) + } + fn retrieve_video(&self, p: &RetrieveVideoParams) -> crate::error::Result { // The Rust `RetrieveVideoParams` carries no `OakRenderRenderer`, so // texture creation cannot be performed — the C++ failure path returns @@ -465,6 +538,108 @@ enum DecodedFrame { Audio(ffmpeg::frame::Audio), } +/// A reference-counted `AVFrame` with `av_frame_ref` clone semantics. +/// +/// ffmpeg-next's `Video::clone` deep-copies with `av_frame_copy`, which +/// does **not** reproduce hardware-surface references (a cloned VAAPI +/// frame loses its buffers and `av_hwframe_transfer_data` then fails with +/// EINVAL). The M5 decode path keeps raw hardware frames in the frame +/// cache, so the cache stores `RefFrame` — cloning one only bumps the +/// buffer references. +pub(crate) struct RefFrame(*mut sys::AVFrame); + +// SAFETY: an `AVFrame` is a plain refcounted buffer holder; FFmpeg +// itself allows moving/freeing frames across threads. Every access is +// serialized by the decoder's state mutex. +unsafe impl Send for RefFrame {} +unsafe impl Sync for RefFrame {} + +impl RefFrame { + /// Take a reference to a raw `AVFrame` (shallow, like + /// [`RefFrame::from_video`]). + pub(crate) fn clone_raw(frame: *const sys::AVFrame) -> Option { + let ptr = unsafe { sys::av_frame_clone(frame) }; + if ptr.is_null() { + None + } else { + Some(RefFrame(ptr)) + } + } + + /// Take a reference to `video`'s `AVFrame` (shallow: buffer refs are + /// incremented, pixel data is not copied). + fn from_video(video: &ffmpeg::frame::Video) -> Option { + let ptr = unsafe { sys::av_frame_clone(video.as_ptr()) }; + if ptr.is_null() { + None + } else { + Some(RefFrame(ptr)) + } + } + + /// A fresh owning `Video` sharing this frame's buffers (the CPU scale + /// path and `av_hwframe_transfer_data` need the owned type). + fn to_video(&self) -> Option { + let video = ffmpeg::frame::Video::empty(); + unsafe { + let dst = video.as_ptr() as *mut sys::AVFrame; + sys::av_frame_unref(dst); + if sys::av_frame_ref(dst, self.0) < 0 { + return None; + } + } + Some(video) + } + + /// The raw frame (borrowed). + fn as_ptr(&self) -> *const sys::AVFrame { + self.0 + } + + /// Presentation timestamp. + fn pts(&self) -> Option { + let pts = unsafe { (*self.0).pts }; + if pts == AV_NOPTS_VALUE { + None + } else { + Some(pts) + } + } + + /// Frame width in pixels. + fn width(&self) -> u32 { + unsafe { (*self.0).width as u32 } + } + + /// Frame height in pixels. + fn height(&self) -> u32 { + unsafe { (*self.0).height as u32 } + } +} + +impl Clone for RefFrame { + fn clone(&self) -> Self { + let ptr = unsafe { sys::av_frame_clone(self.0) }; + assert!(!ptr.is_null(), "av_frame_clone failed"); + RefFrame(ptr) + } +} + +impl Drop for RefFrame { + fn drop(&mut self) { + unsafe { + let mut ptr = self.0; + sys::av_frame_free(&mut ptr); + } + } +} + +impl std::fmt::Debug for RefFrame { + fn fmt(&self, f: &mut std::fmt::Formatter<'_>) -> std::fmt::Result { + f.debug_struct("RefFrame").field("pts", &self.pts()).finish() + } +} + /// Outcome of a single decoder receive attempt. enum Pull { Frame(DecodedFrame), @@ -521,7 +696,7 @@ unsafe impl Sync for DecoderState {} /// Video decode state: the frame cache plus a cached swscale context. struct VideoDecodeState { scaler: Option, - cache: VecDeque, + cache: VecDeque, cache_at_zero: bool, cache_at_eof: bool, /// One second in the stream's time base. @@ -895,7 +1070,7 @@ impl DecoderState { time: &Rational, any_timecode: bool, cancelled: Option<&CancelAtom>, - ) -> crate::error::Result> { + ) -> crate::error::Result> { // Move the video state out so `self.seek` / `self.pull` (which touch // other fields) can be called without conflicting borrows. let mut video = self @@ -942,7 +1117,7 @@ impl DecoderState { } let mut retried_after_eof = false; - let mut return_frame: Option = None; + let mut return_frame: Option = None; loop { if cancel_atom_is_cancelled(cancelled) { @@ -951,20 +1126,13 @@ impl DecoderState { let frame = match self.pull()? { Pull::Frame(DecodedFrame::Video(f)) => { - // A hardware decoder yields hardware surfaces - // (AV_PIX_FMT_VIDEOTOOLBOX/VAAPI/CUDA/D3D11*): - // transfer to system memory so the cache and swscale - // only ever see CPU frames. - // SAFETY: plain read of the frame's format field. - let raw_format = unsafe { (*f.as_ptr()).format }; - if self.hw_device.is_some() - && crate::hwdecode::is_hw_format(unsafe { - std::mem::transmute::(raw_format) - }) { - crate::hwdecode::transfer_to_cpu(&f)? - } else { - f - } + // Hardware surfaces stay as they are: the import path + // (M5) and `scale_video_to_f32` decide between + // zero-copy GPU import and the CPU transfer. Keeping + // the raw frame in the cache lets an import attempt + // and a CPU fallback share one decode. + RefFrame::from_video(&f) + .ok_or_else(|| fail("frame reference allocation failed"))? } Pull::Frame(_) => unreachable!("video session yields only video frames"), Pull::Eof => { @@ -1070,47 +1238,39 @@ impl DecoderState { /// full-resolution float intermediate (~132 MB at 4K) ever exists. fn scale_video_to_f32( &mut self, - f: ffmpeg::frame::Video, + f: RefFrame, force_range: i32, target_size: Option<(u32, u32)>, ) -> crate::error::Result<(u32, u32, Vec, crate::decoder::DecodedColorMeta)> { + // The frame's own colorimetry (set by the decoder from the + // bitstream); raw code points pass through to the render layer. + let (raw_primaries, raw_trc, raw_space, full_range) = + frame_colorimetry(f.as_ptr(), force_range); + let f = f + .to_video() + .ok_or_else(|| fail("frame reference failed"))?; + // Hardware frames reaching the CPU path are downloaded here — the + // per-frame staging fallback when the zero-copy import declined + // the frame (or the caller explicitly wants CPU pixels). + let f = if crate::hwdecode::is_hw_format(unsafe { + std::mem::transmute::((*f.as_ptr()).format) + }) { + crate::hwdecode::transfer_to_cpu(&f)? + } else { + f + }; let video = self .video .as_mut() .expect("scale_video_to_f32 requires a video session"); - // The frame's own colorimetry (set by the decoder from the - // bitstream); raw code points pass through to the render layer. - let (raw_primaries, raw_trc, raw_space, raw_range) = unsafe { - let av = f.as_ptr(); - ( - (*av).color_primaries as i32, - (*av).color_trc as i32, - (*av).colorspace as i32, - (*av).color_range as i32, - ) - }; - // # CPP-PARITY ffmpegdecoder.cpp:376: disregard "JPEG" pixel formats // — but a YUVJ source is full range by definition, so remember it // for the range decision below. let orig_format = f.format(); let src_format = convert_jpeg_space_to_regular_space(orig_format); - let yuvj_full = orig_format != src_format; let mut f = f; f.set_format(src_format); - - // The effective color range: the caller's force wins; otherwise the - // frame's own metadata (YUVJ sources are full range). The old path - // forced MPEG/limited for everything, crushing full-range screen - // captures and JPEG-derived footage. - let full_range = if force_range == oak_core_COLOR_RANGE_FULL { - true - } else if force_range == oak_core_COLOR_RANGE_LIMITED { - false - } else { - yuvj_full || raw_range == AVCOL_RANGE_JPEG - }; f.set_color_range(if full_range { ffmpeg::color::Range::JPEG } else { @@ -1800,6 +1960,36 @@ fn yuv_matrix_for(av_colorspace: i32, src_w: u32, src_h: u32) -> YuvMatrix { } } +/// The frame's raw colorimetry and effective range: `(color_primaries, +/// color_trc, colorspace, full_range)`. Shared by the CPU scale path and +/// the GPU import path so both hand the render layer identical +/// colorimetry. `force_range` wins when it is one of the explicit +/// `oak_core_COLOR_RANGE_*` values; otherwise the frame's own metadata +/// decides (YUVJ sources are full range by definition — the caller +/// resolves the format before this call for the CPU path, but the raw +/// AVFrame format is also checked here). +fn frame_colorimetry(f: *const sys::AVFrame, force_range: i32) -> (i32, i32, i32, bool) { + // SAFETY: plain reads of the frame's color fields. + let (primaries, trc, space, range) = unsafe { + ( + (*f).color_primaries as i32, + (*f).color_trc as i32, + (*f).colorspace as i32, + (*f).color_range as i32, + ) + }; + let format = unsafe { Pixel::from(std::mem::transmute::((*f).format)) }; + let yuvj_full = convert_jpeg_space_to_regular_space(format) != format; + let full_range = if force_range == oak_core_COLOR_RANGE_FULL { + true + } else if force_range == oak_core_COLOR_RANGE_LIMITED { + false + } else { + yuvj_full || range == AVCOL_RANGE_JPEG + }; + (primaries, trc, space, full_range) +} + /// Bit depth (bits per component) and YUV-ness of a pixel format, from its /// `AVPixFmtDescriptor` (8 and false for formats without one — none in /// practice for decoder output). @@ -1845,7 +2035,7 @@ fn apply_sws_colorspace( } /// The frame's presentation timestamp (NOPTS when unset). -fn pts_of(f: Option<&ffmpeg::frame::Video>) -> Option { +fn pts_of(f: Option<&RefFrame>) -> Option { f.and_then(|f| f.pts()) } @@ -1853,7 +2043,7 @@ fn pts_of(f: Option<&ffmpeg::frame::Video>) -> Option { /// /// # CPP-PARITY /// `FFmpegDecoder::get_frame_from_cache`. -fn get_frame_from_cache(video: &VideoDecodeState, t: i64) -> Option { +fn get_frame_from_cache(video: &VideoDecodeState, t: i64) -> Option { let front = pts_of(video.cache.front()).unwrap_or(AV_NOPTS_VALUE); let back = pts_of(video.cache.back()).unwrap_or(AV_NOPTS_VALUE); if t < front { diff --git a/crates/oak-codec/src/gpuinterop.rs b/crates/oak-codec/src/gpuinterop.rs new file mode 100644 index 000000000..56f01645e --- /dev/null +++ b/crates/oak-codec/src/gpuinterop.rs @@ -0,0 +1,715 @@ +// Oak Video Editor - Non-Linear Video Editor +// Copyright (C) 2026 Oak Team +// +// This program is free software: you can redistribute it and/or modify +// it under the terms of the GNU General Public License as published by +// the Free Software Foundation, either version 3 of the License, or +// (at your option) any later version. +// +// This program is distributed in the hope that it will be useful, +// but WITHOUT ANY WARRANTY; without even the implied warranty of +// MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. See the +// GNU General Public License for more details. +// +// You should have received a copy of the GNU General Public License +// along with this program. If not, see . + +//! M5: zero-copy hardware-decode import. +//! +//! A decoded hardware surface (VAAPI/NVDEC/D3D11VA/VideoToolbox) is +//! imported into the engine's wgpu device as a pair of planar textures +//! (luma + interleaved chroma) instead of being downloaded with +//! `av_hwframe_transfer_data`. The render side turns the planes into +//! working-space RGBA with the M2 YUV→RGB pass; the CPU staging path +//! stays the per-frame fallback. +//! +//! Platform surface extraction lives here (FFmpeg types); the raw +//! Vulkan/Metal import lives in `oak_core::backend::external`. +//! +//! ## Counters +//! +//! - [`HW_IMPORTS`] — frames that took the zero-copy path. +//! - [`HW_IMPORT_FALLBACKS`] — attempts that were tried and failed. +//! - [`HW_IMPORT_UNSUPPORTED`] — frames that skipped the import because +//! no import path exists (NVDEC, a platform switch off, an +//! unimportable layout); not failures. +//! - `hwdecode::HW_TRANSFERS` — frames downloaded to system memory; a +//! zero-copy hardware frame must not move it. + +use std::sync::Arc; +use std::sync::atomic::{AtomicU32, AtomicU64, Ordering}; + +use ffmpeg::ffi as sys; +use ffmpeg_next as ffmpeg; +use oak_core::backend::{GpuContext, YuvTransform}; +#[cfg(target_os = "linux")] +use oak_core::backend::ImportGuard; +use oak_core::colormath::YuvMatrix; +use oak_core::texture::PlanarFormat; + +/// Frames that were imported to the GPU instead of downloaded (M5 +/// acceptance: the hardware path keeps `hwdecode::HW_TRANSFERS` at 0). +pub static HW_IMPORTS: AtomicU64 = AtomicU64::new(0); + +/// Import attempts that were actually attempted and failed (per-frame +/// fallback; feeds the consecutive-failure breaker). "Unsupported" +/// outcomes are deliberately NOT counted here — a missing import path is +/// not a failure (M5 audit). +pub static HW_IMPORT_FALLBACKS: AtomicU64 = AtomicU64::new(0); + +/// Frames that took the staging path because no import path exists for +/// this device/machine/configuration: NVDEC/CUDA has no public handle +/// export, a per-platform switch is off, or the surface layout is not +/// importable. Observability only: neither counted as a fallback nor fed +/// to the breaker. +pub static HW_IMPORT_UNSUPPORTED: AtomicU64 = AtomicU64::new(0); + +/// Consecutive import failures after which the process stops attempting +/// (the platform combination is evidently unsupported on this machine). +/// A single failure does not disable anything — the next frame is +/// attempted afresh, as the per-frame fallback contract requires. +const STICKY_FAILURE_LIMIT: u32 = 8; + +static CONSECUTIVE_FAILURES: AtomicU32 = AtomicU32::new(0); + +/// The config key of the zero-copy import switch (1 = import, 0 = +/// staging). Default ON; `OAK_GPU_IMPORT=0` overrides both. +pub const CONFIG_KEY_GPU_DECODE_IMPORT: &str = "GpuDecodeImport"; + +/// The per-platform import switches (each platform's import is an +/// independent, revertable row of design §3.6). +pub const CONFIG_KEY_VAAPI_IMPORT: &str = "GpuDecodeImportVaapi"; +/// Windows D3D11VA import switch. +pub const CONFIG_KEY_D3D11_IMPORT: &str = "GpuDecodeImportD3d11"; +/// macOS VideoToolbox import switch. +pub const CONFIG_KEY_VIDEOTOOLBOX_IMPORT: &str = "GpuDecodeImportVideotoolbox"; + +/// Whether the zero-copy hardware import is preferred. `OAK_GPU_IMPORT=0` +/// force-disables it (diagnostics/CI without importable surfaces), same +/// convention as `OAK_HWACCEL`. +pub fn gpu_import_enabled() -> bool { + if let Ok(v) = std::env::var("OAK_GPU_IMPORT") { + return v != "0"; + } + switch_from_config(CONFIG_KEY_GPU_DECODE_IMPORT) +} + +/// One platform's import switch (global import switch AND platform +/// switch). `env` overrides the config key, like the global switch. +fn platform_import_enabled(env: &str, config: &str) -> bool { + if let Ok(v) = std::env::var(env) { + return v != "0"; + } + switch_from_config(config) +} + +fn switch_from_config(key: &str) -> bool { + match oak_core::configstore::ConfigStore::instance().get(None, key) { + Ok(value) => value != "false", + Err(_) => true, + } +} + +/// Reset the import counters (tests). +pub fn reset_import_counters() { + HW_IMPORTS.store(0, Ordering::Relaxed); + HW_IMPORT_FALLBACKS.store(0, Ordering::Relaxed); + HW_IMPORT_UNSUPPORTED.store(0, Ordering::Relaxed); + CONSECUTIVE_FAILURES.store(0, Ordering::Relaxed); +} + +/// Everything the import needs from the decoder's frame and colorimetry. +pub struct HwImportRequest<'a> { + /// The decoded frame (a hardware surface when import is possible). + /// Borrowed for the duration of the attempt; the import guard takes + /// its own reference to the underlying surface. + pub frame: *const sys::AVFrame, + /// Lifetime marker for the borrowed frame. + pub _marker: std::marker::PhantomData<&'a ()>, + /// The session's hardware device type (`None` = software decode). + pub device_type: Option, + /// Frame dimensions (luma). + pub luma_size: (u32, u32), + /// The frame's YUV→RGB luma matrix. + pub matrix: YuvMatrix, + /// Full/limited range. + pub full_range: bool, + /// Source color-primaries code point. + pub color_primaries: i32, + /// Source transfer-characteristic code point. + pub color_trc: i32, +} + +/// A successfully imported hardware frame: the plane tokens plus the +/// colorimetry the render side needs to resolve them to working-space +/// RGBA. +pub struct ImportedHwFrame { + /// Luma plane token. + pub y: u64, + /// Interleaved chroma plane token. + pub uv: u64, + /// Plane layout (bit depth / interleave). + pub format: PlanarFormat, + /// Luma dimensions. + pub width: u32, + /// Frame height. + pub height: u32, + /// The GPU device that owns the planes. + pub ctx: Arc, + /// Extra lifetime guard for platforms whose GPU texture types cannot + /// carry a drop callback (macOS): the decoder's pixel buffer. + pub keep_alive: Option>, +} + +impl ImportedHwFrame { + /// The YUV→RGB transform for the planes (the bit depth picks the + /// exact limited-range expansion: 8 for NV12, 16 for P010). + pub fn transform(&self, matrix: YuvMatrix, full_range: bool) -> YuvTransform { + YuvTransform::from_matrix_depth(matrix, full_range, self.format.bit_depth()) + } +} + +/// Try to import `req.frame` as a pair of planar GPU textures. `None` +/// means "use the staging path for this frame" — the caller must then +/// call the CPU retrieval (the decoder session cache still holds the +/// decoded surface, so no re-decode happens). +pub fn try_import_hw_frame(req: &HwImportRequest<'_>) -> Option { + if !gpu_import_enabled() { + return None; + } + if CONSECUTIVE_FAILURES.load(Ordering::Relaxed) >= STICKY_FAILURE_LIMIT { + return None; + } + let device_type = req.device_type?; + // Only a host-installed device (the app's UI/render context) may take + // the frame: a lazily created worker/CLI context must keep decoding on + // the CPU path. Checked before touching the frame so the common + // "no import here" case costs nothing. + if !GpuContext::host_gpu_installed() { + return None; + } + if !crate::hwdecode::is_hw_format(frame_format(req.frame)) { + return None; + } + let Some(ctx) = GpuContext::shared() else { + // The host slot was cleared between the two checks (shutdown): + // nothing was attempted, so this is not a failure. + HW_IMPORT_UNSUPPORTED.fetch_add(1, Ordering::Relaxed); + return None; + }; + match platform_import(req, &ctx, device_type) { + ImportOutcome::Imported { + y, + uv, + format, + keep_alive, + } => { + CONSECUTIVE_FAILURES.store(0, Ordering::Relaxed); + HW_IMPORTS.fetch_add(1, Ordering::Relaxed); + Some(ImportedHwFrame { + y, + uv, + format, + width: req.luma_size.0, + height: req.luma_size.1, + ctx, + keep_alive, + }) + } + ImportOutcome::Unsupported => { + HW_IMPORT_UNSUPPORTED.fetch_add(1, Ordering::Relaxed); + None + } + ImportOutcome::Failed => { + bump_fallback(); + None + } + } +} + +/// Outcome of one platform import attempt. +enum ImportOutcome { + /// The planes were imported; the tokens and keep-alive guard. + Imported { + y: u64, + uv: u64, + format: PlanarFormat, + keep_alive: Option>, + }, + /// No import path exists for this device/machine/configuration (the + /// frame takes the staging path without being a failure). + Unsupported, + /// The import path exists and the attempt failed: per-frame fallback, + /// counted and fed to the consecutive-failure breaker. + Failed, +} + +fn bump_fallback() { + HW_IMPORT_FALLBACKS.fetch_add(1, Ordering::Relaxed); + CONSECUTIVE_FAILURES.fetch_add(1, Ordering::Relaxed); +} + +/// The frame's raw pixel format (hardware variants included). +fn frame_format(frame: *const sys::AVFrame) -> sys::AVPixelFormat { + // SAFETY: plain read of the frame's format field. + unsafe { std::mem::transmute::((*frame).format) } +} + +#[cfg(target_os = "linux")] +fn platform_import( + req: &HwImportRequest<'_>, + ctx: &Arc, + device_type: sys::AVHWDeviceType, +) -> ImportOutcome { + use sys::AVHWDeviceType; + if device_type != AVHWDeviceType::AV_HWDEVICE_TYPE_VAAPI { + // CUDA/NVDEC surfaces have no public fd/handle export; the + // staging path is the honest fallback (design §3.6) — NOT a + // failure. + return ImportOutcome::Unsupported; + } + if !platform_import_enabled("OAK_GPU_IMPORT_VAAPI", CONFIG_KEY_VAAPI_IMPORT) { + return ImportOutcome::Unsupported; + } + vaapi_import(req, ctx) +} + +#[cfg(target_os = "windows")] +fn platform_import( + req: &HwImportRequest<'_>, + ctx: &Arc, + device_type: sys::AVHWDeviceType, +) -> ImportOutcome { + if device_type != sys::AVHWDeviceType::AV_HWDEVICE_TYPE_D3D11VA { + return ImportOutcome::Unsupported; + } + if !platform_import_enabled("OAK_GPU_IMPORT_D3D11", CONFIG_KEY_D3D11_IMPORT) { + return ImportOutcome::Unsupported; + } + d3d11_import(req, ctx) +} + +#[cfg(target_os = "macos")] +fn platform_import( + req: &HwImportRequest<'_>, + ctx: &Arc, + device_type: sys::AVHWDeviceType, +) -> ImportOutcome { + if device_type != sys::AVHWDeviceType::AV_HWDEVICE_TYPE_VIDEOTOOLBOX { + return ImportOutcome::Unsupported; + } + if !platform_import_enabled("OAK_GPU_IMPORT_VIDEOTOOLBOX", CONFIG_KEY_VIDEOTOOLBOX_IMPORT) { + return ImportOutcome::Unsupported; + } + videotoolbox_import(req, ctx) +} + +#[cfg(not(any(target_os = "linux", target_os = "windows", target_os = "macos")))] +fn platform_import( + _req: &HwImportRequest<'_>, + _ctx: &Arc, + _device_type: sys::AVHWDeviceType, +) -> ImportOutcome { + ImportOutcome::Unsupported +} + +// --------------------------------------------------------------------------- +// Linux: VAAPI surface -> DRM PRIME descriptor -> DMA-BUF plane imports +// --------------------------------------------------------------------------- + +/// DRM format fourcc values (little-endian `fourcc_code`). +#[cfg(target_os = "linux")] +mod drm_fourcc { + pub const R8: u32 = 0x2020_3852; // 'R','8',' ',' ' + pub const R16: u32 = 0x2036_3152; // 'R','1','6',' ' + pub const GR88: u32 = 0x3838_5252; // 'R','R','8','8' + pub const RG88: u32 = 0x3838_4752; // 'R','G','8','8' + pub const RG1616: u32 = 0x3631_4752; // 'R','G','1','6' +} + + +/// Map a VAAPI frame to its DRM PRIME descriptor (FFmpeg calls +/// `vaExportSurfaceHandle` internally; no CPU download happens). +#[cfg(target_os = "linux")] +fn vaapi_import(req: &HwImportRequest<'_>, ctx: &Arc) -> ImportOutcome { + let mut drm_frame = ffmpeg::frame::Video::empty(); + // SAFETY: `drm_frame` is a valid AVFrame; the map only writes into it + // and takes references to the source frame's hwframe mapping. + let rc = unsafe { + let raw = drm_frame.as_mut_ptr(); + (*raw).format = sys::AVPixelFormat::AV_PIX_FMT_DRM_PRIME as i32; + sys::av_hwframe_map(raw, req.frame, sys::AV_HWFRAME_MAP_READ as i32) + }; + if rc < 0 { + // The VAAPI session exists but the surface could not be exported + // as DRM PRIME: a real failure (the breaker bounds the retries). + return ImportOutcome::Failed; + } + // SAFETY: on success `data[0]` is the descriptor FFmpeg allocated for + // this mapping; it stays valid until `drm_frame` is unref'd. + let desc = unsafe { &*((*drm_frame.as_ptr()).data[0] as *const sys::AVDRMFrameDescriptor) }; + if desc.nb_layers != 2 { + return ImportOutcome::Unsupported; + } + let layer_format = |i: usize| desc.layers[i].format; + let (format, y_fourcc, uv_fourcc) = match (layer_format(0), layer_format(1)) { + (a, b) if a == drm_fourcc::R8 && (b == drm_fourcc::GR88 || b == drm_fourcc::RG88) => { + (PlanarFormat::Nv12, a, b) + } + (a, b) if a == drm_fourcc::R16 && b == drm_fourcc::RG1616 => { + (PlanarFormat::P010, a, b) + } + _ => return ImportOutcome::Unsupported, + }; + let _ = (y_fourcc, uv_fourcc); + for layer in &desc.layers[..desc.nb_layers as usize] { + if layer.nb_planes != 1 { + return ImportOutcome::Unsupported; + } + } + let (lw, lh) = req.luma_size; + if lw == 0 || lh == 0 { + return ImportOutcome::Unsupported; + } + let (cw, ch) = (lw.div_ceil(2), lh.div_ceil(2)); + let plane = |layer: usize| { + let l = &desc.layers[layer]; + let p = &l.planes[0]; + let obj = &desc.objects[p.object_index as usize]; + oak_core::backend::external::DmaBufPlane { + fd: obj.fd, + offset: p.offset as u64, + pitch: p.pitch as u64, + modifier: obj.format_modifier, + } + }; + // Shared keep-alive: the source frame pins the VAAPI surface until + // the last plane texture is destroyed (the dmabuf fds Vulkan imported + // keep the buffer itself alive even after the mapping drops). + let Some(keep) = crate::ffmpeg::RefFrame::clone_raw(req.frame) else { + return ImportOutcome::Failed; + }; + let keep = Arc::new(keep); + let y_guard = keep.clone(); + let y = match ctx.import_dmabuf_plane(lw, lh, format.y_format(), plane(0), guard(y_guard)) { + Ok(y) => y, + Err(oak_core::error::Error::State) | Err(oak_core::error::Error::Invalid) => { + return ImportOutcome::Unsupported + } + Err(_) => return ImportOutcome::Failed, + }; + let uv_guard = keep.clone(); + match ctx.import_dmabuf_plane(cw, ch, format.uv_format(), plane(1), guard(uv_guard)) { + Ok(uv) => ImportOutcome::Imported { + y, + uv, + format, + keep_alive: None, + }, + Err(_) => { + ctx.destroy_texture(y); + ImportOutcome::Failed + } + } +} + +/// A keep-alive import guard holding one `Arc` of the decoder frame +/// (Linux/VAAPI path; Windows attaches its guard inline and macOS keeps +/// the frame on the planar texture). +#[cfg(target_os = "linux")] +fn guard(value: Arc) -> Option { + Some(Box::new(move || drop(value))) +} + +// --------------------------------------------------------------------------- +// Windows: D3D11VA texture -> shared NT handle -> Vulkan import +// --------------------------------------------------------------------------- + +/// Import a D3D11VA texture (Windows): one NV12/P010 texture shared as an +/// NT handle, imported as a single multi-planar image whose two plane +/// views feed the planar YUV→RGB pass. The decoder's frame reference is +/// kept alive by the texture's drop guard. +#[cfg(target_os = "windows")] +fn d3d11_import(req: &HwImportRequest<'_>, ctx: &Arc) -> ImportOutcome { + use windows::core::Interface; + use windows::Win32::Foundation::CloseHandle; + use windows::Win32::Graphics::Direct3D11::{ID3D11Texture2D, D3D11_TEXTURE2D_DESC}; + use windows::Win32::Graphics::Dxgi::Common::{DXGI_FORMAT_NV12, DXGI_FORMAT_P010}; + use windows::Win32::Graphics::Dxgi::{IDXGIResource1, DXGI_SHARED_RESOURCE_READ}; + + // AV_PIX_FMT_D3D11: data[0] = ID3D11Texture2D*, data[1] = subresource + // index. D3D11VA frames cycle through the decoder's surface ARRAY, so + // the slice index is part of the frame identity — the import selects + // it via an array-layer view (M5 audit). + let texture_ptr = unsafe { (*req.frame).data[0] } as *mut std::ffi::c_void; + if texture_ptr.is_null() { + return ImportOutcome::Unsupported; + } + let slice = unsafe { (*req.frame).data[1] } as usize as u32; + // SAFETY: FFmpeg owns the texture reference for the frame's lifetime. + let texture: &ID3D11Texture2D = unsafe { &*(texture_ptr as *const ID3D11Texture2D) }; + let mut desc = D3D11_TEXTURE2D_DESC::default(); + unsafe { texture.GetDesc(&mut desc) }; + let array_size = desc.ArraySize.max(1); + if slice >= array_size { + return ImportOutcome::Unsupported; + } + let (texture_format, planar) = match desc.Format { + DXGI_FORMAT_NV12 => (oak_core::wgpu::TextureFormat::NV12, PlanarFormat::Nv12), + DXGI_FORMAT_P010 => (oak_core::wgpu::TextureFormat::P010, PlanarFormat::P010), + _ => return ImportOutcome::Unsupported, + }; + // SAFETY: the D3D11 texture supports IDXGIResource1 (every shared + // texture does); a cast failure means it is not a shareable resource. + let Ok(resource) = texture.cast::() else { + return ImportOutcome::Unsupported; + }; + // Read-only NT handle; Vulkan references the resource, so the handle + // is closed again right after the import. + let handle = unsafe { + resource.CreateSharedHandle( + None, + DXGI_SHARED_RESOURCE_READ.0, + None::<&windows::core::PCWSTR>, + ) + }; + let Ok(handle) = handle else { + return ImportOutcome::Failed; + }; + let Some(keep) = crate::ffmpeg::RefFrame::clone_raw(req.frame) else { + unsafe { + let _ = CloseHandle(handle); + } + return ImportOutcome::Failed; + }; + let (w, h) = req.luma_size; + let guard: Arc = Arc::new(keep); + let result = ctx.import_d3d11_shared_texture( + handle.0 as isize, + w, + h, + texture_format, + array_size, + slice, + Some(Box::new(move || drop(guard))), + ); + unsafe { + let _ = CloseHandle(handle); + } + match result { + Ok((y, uv)) => ImportOutcome::Imported { + y, + uv, + format: planar, + keep_alive: None, + }, + // State/Invalid are capability gaps (missing device feature, wrong + // format), not failed attempts. + Err(oak_core::error::Error::State) | Err(oak_core::error::Error::Invalid) => { + ImportOutcome::Unsupported + } + Err(_) => ImportOutcome::Failed, + } +} + +// --------------------------------------------------------------------------- +// macOS: VideoToolbox CVPixelBuffer -> IOSurface -> Metal textures +// --------------------------------------------------------------------------- + +/// `CVPixelBufferGetIOSurface` (CoreVideo). +#[cfg(target_os = "macos")] +#[link(name = "CoreVideo", kind = "framework")] +extern "C" { + fn CVPixelBufferGetIOSurface(pixel_buffer: *mut std::ffi::c_void) -> *mut std::ffi::c_void; +} + +/// Import a VideoToolbox frame (macOS): its CVPixelBuffer's IOSurface is +/// wrapped in one Metal texture per plane. Metal textures cannot carry a +/// drop callback, so the decoder's frame reference travels back in +/// `ImportedHwFrame::keep_alive` and is held by the planar texture. +#[cfg(target_os = "macos")] +fn videotoolbox_import(req: &HwImportRequest<'_>, ctx: &Arc) -> ImportOutcome { + // AV_PIX_FMT_VIDEOTOOLBOX keeps the CVPixelBuffer in data[3]. + let pixel_buffer = unsafe { (*req.frame).data[3] } as *mut std::ffi::c_void; + if pixel_buffer.is_null() { + return ImportOutcome::Unsupported; + } + let surface = unsafe { CVPixelBufferGetIOSurface(pixel_buffer) }; + if surface.is_null() { + return ImportOutcome::Unsupported; + } + let frames = unsafe { (*req.frame).hw_frames_ctx }; + if frames.is_null() { + return ImportOutcome::Unsupported; + } + // SAFETY: `hw_frames_ctx` is an AVHWFramesContext for this frame. + let sw_format = unsafe { (*(*frames).data as *const sys::AVHWFramesContext).sw_format }; + let format = match sw_format { + sys::AVPixelFormat::AV_PIX_FMT_NV12 => PlanarFormat::Nv12, + sys::AVPixelFormat::AV_PIX_FMT_P010LE | sys::AVPixelFormat::AV_PIX_FMT_P010BE => { + PlanarFormat::P010 + } + _ => return ImportOutcome::Unsupported, + }; + let (w, h) = req.luma_size; + let (cw, ch) = (w.div_ceil(2), h.div_ceil(2)); + // SAFETY: `surface` is the CVPixelBuffer's IOSurface and the decoder + // frame reference is kept alive by the import guard (planar texture). + let Ok(y) = (unsafe { ctx.import_iosurface_plane(surface, 0, w, h, format.y_format()) }) else { + return ImportOutcome::Failed; + }; + let uv = match unsafe { ctx.import_iosurface_plane(surface, 1, cw, ch, format.uv_format()) } { + Ok(uv) => uv, + Err(_) => { + ctx.destroy_texture(y); + return ImportOutcome::Failed; + } + }; + let Some(keep) = crate::ffmpeg::RefFrame::clone_raw(req.frame) else { + ctx.destroy_texture(y); + ctx.destroy_texture(uv); + return ImportOutcome::Failed; + }; + ImportOutcome::Imported { + y, + uv, + format, + keep_alive: Some(Arc::new(keep)), + } +} + +#[cfg(test)] +mod tests { + use super::*; + + fn empty_frame() -> ffmpeg::frame::Video { + ffmpeg::frame::Video::empty() + } + + fn request(frame: &ffmpeg::frame::Video, device_type: sys::AVHWDeviceType) -> HwImportRequest<'_> { + HwImportRequest { + frame: unsafe { frame.as_ptr() }, + _marker: std::marker::PhantomData, + device_type: Some(device_type), + luma_size: (64, 64), + matrix: YuvMatrix::Bt709, + full_range: false, + color_primaries: 1, + color_trc: 1, + } + } + + /// Test-only RAII override for a process-global environment variable: + /// restores the previous value (or absence) on drop, so a panicking + /// assertion cannot leak the override into the next serialized test and + /// an externally preset value survives the test. + struct EnvVarGuard { + name: &'static str, + previous: Option, + } + + impl EnvVarGuard { + /// Set `name=value` for the guard's lifetime. + fn set(name: &'static str, value: &str) -> Self { + let previous = std::env::var_os(name); + std::env::set_var(name, value); + Self { name, previous } + } + + /// Remove `name` for the guard's lifetime. + fn remove(name: &'static str) -> Self { + let previous = std::env::var_os(name); + std::env::remove_var(name); + Self { name, previous } + } + } + + impl Drop for EnvVarGuard { + fn drop(&mut self) { + match self.previous.take() { + Some(value) => std::env::set_var(self.name, value), + None => std::env::remove_var(self.name), + } + } + } + + /// Without a host-installed device the import declines before any + /// frame work (and is not counted as a fallback: nothing failed). + #[test] + fn import_requires_a_host_installed_context() { + let _g = crate::lock_tests(); + reset_import_counters(); + assert!(!GpuContext::host_gpu_installed()); + let frame = empty_frame(); + let req = request(&frame, sys::AVHWDeviceType::AV_HWDEVICE_TYPE_VAAPI); + assert!(try_import_hw_frame(&req).is_none()); + assert_eq!(HW_IMPORTS.load(Ordering::Relaxed), 0); + assert_eq!(HW_IMPORT_FALLBACKS.load(Ordering::Relaxed), 0); + assert_eq!(HW_IMPORT_UNSUPPORTED.load(Ordering::Relaxed), 0); + } + + /// The counter classification (M5 audit): a device type with no import + /// path (NVDEC/CUDA) is "unsupported", never a failure — it must not + /// feed the consecutive-failure breaker or `HW_IMPORT_FALLBACKS`. + /// Exercised at the platform-function level so no real GPU is needed. + /// + /// (`platform_import` is private; this test lives in the same module.) + #[cfg(target_os = "linux")] + #[test] + fn non_importable_device_type_is_unsupported_not_failed() { + let _g = crate::lock_tests(); + // A context is required by the platform function's signature; + // the non-VAAPI device check happens before any of its fields are + // touched, so a lazily-created local context is never used. + let Some(ctx) = GpuContext::create(oak_core::backend::BackendKind::Auto) else { + eprintln!("no GPU adapter; skipping classification test"); + return; + }; + let frame = empty_frame(); + let req = request(&frame, sys::AVHWDeviceType::AV_HWDEVICE_TYPE_CUDA); + assert!( + matches!( + platform_import(&req, &ctx, sys::AVHWDeviceType::AV_HWDEVICE_TYPE_CUDA), + ImportOutcome::Unsupported + ), + "CUDA/NVDEC has no export path and must classify as unsupported" + ); + } + + /// The `OAK_GPU_IMPORT=0` escape hatch disables the path outright, and + /// removing the override falls back to the default-on config. The + /// override is RAII-scoped, so an externally preset value survives the + /// test instead of being clobbered by `remove_var`. + #[test] + fn import_switch_disables_the_path() { + let _g = crate::lock_tests(); + let previous = std::env::var_os("OAK_GPU_IMPORT"); + { + let _off = EnvVarGuard::set("OAK_GPU_IMPORT", "0"); + assert!(!gpu_import_enabled(), "explicit 0 disables the import path"); + } + assert_eq!( + std::env::var_os("OAK_GPU_IMPORT"), + previous, + "the guard restores the externally preset value" + ); + if previous.is_none() { + // No override: the config-store default keeps the import path + // on. (With an external value the caller's setting governs, so + // asserting the default would be wrong.) + let _unset = EnvVarGuard::remove("OAK_GPU_IMPORT"); + assert!( + gpu_import_enabled(), + "unset falls back to the default-on config" + ); + } + } + + /// Plane layouts carry the expansion depth the YUV transform needs. + #[test] + fn planar_formats_carry_their_bit_depth() { + assert_eq!(PlanarFormat::Nv12.bit_depth(), 8); + assert_eq!(PlanarFormat::P010.bit_depth(), 16); + assert_eq!(PlanarFormat::Nv12.name(), "nv12"); + } +} diff --git a/crates/oak-codec/src/hwdecode.rs b/crates/oak-codec/src/hwdecode.rs index f4e0d185e..14d76c356 100644 --- a/crates/oak-codec/src/hwdecode.rs +++ b/crates/oak-codec/src/hwdecode.rs @@ -31,14 +31,15 @@ //! hardware surface. //! //! - **macOS**: `AV_HWDEVICE_TYPE_VIDEOTOOLBOX` -//! - **Linux**: `CUDA` (NVDEC, the discrete-GPU path), then `VAAPI` +//! - **Linux**: `VAAPI` (DMA-BUF, the M5 import path), then `CUDA` (NVDEC) //! - **Windows**: `D3D11VA`, then `CUDA` (NVDEC) //! //! Device creation can fail on machines without the device/driver (a //! headless Linux box, no NVIDIA GPU) — the candidate is skipped and //! the next one (or the software decoder) is used. Hardware frames //! (`AV_PIX_FMT_VIDEOTOOLBOX` / `VAAPI` / `CUDA` / `D3D11VA_VLD` / -//! `D3D11`) are transferred to system memory with +//! `D3D11`) either go through the M5 zero-copy import +//! (`crate::gpuinterop`) or are transferred to system memory with //! `av_hwframe_transfer_data` before the swscale conversion. use ffmpeg::ffi as sys; @@ -131,16 +132,17 @@ pub fn device_type_candidates() -> &'static [sys::AVHWDeviceType] { } #[cfg(all(unix, not(target_os = "macos")))] { - // CUDA (native NVDEC) first: it is the discrete-GPU path, and on - // multi-GPU boxes VAAPI's default render node can point at a - // device with no VA driver (e.g. NVIDIA without - // libva-nvidia-driver) while the AMD/Intel node would work — - // trying NVDEC first sidesteps that misdirection. VAAPI stays as - // the fallback for AMD/Intel-only machines (CUDA device creation - // fails fast without libcuda). + // VAAPI first: it is the DMA-BUF path the M5 zero-copy import + // consumes (`VK_EXT_external_memory_dma_buf`), and on NVIDIA boxes + // with the VA-API driver it is just as reachable as NVDEC — while + // a CUDA (NVDEC) surface has no public handle export, so frames + // decoded through it can only take the staging fallback. CUDA + // stays as the fallback for machines without libva; it also + // sidesteps VAAPI's default render node pointing at a device with + // no VA driver (CUDA device creation fails fast without libcuda). &[ - sys::AVHWDeviceType::AV_HWDEVICE_TYPE_CUDA, sys::AVHWDeviceType::AV_HWDEVICE_TYPE_VAAPI, + sys::AVHWDeviceType::AV_HWDEVICE_TYPE_CUDA, ] } } @@ -363,3 +365,6 @@ mod tests { assert!(!device_unavailable(dev)); } } ++ /// would let their attempts perturb the negative-cache assertion on ++#[cfg(test)] ++#[cfg(test)] diff --git a/crates/oak-codec/src/lib.rs b/crates/oak-codec/src/lib.rs index 865476335..a68cbc6ca 100644 --- a/crates/oak-codec/src/lib.rs +++ b/crates/oak-codec/src/lib.rs @@ -46,6 +46,7 @@ pub mod ffmpeg; pub mod footagedescription; pub mod frame; pub mod framemanager; +pub mod gpuinterop; pub mod hwdecode; pub mod oiio; pub mod oiioframebridge; diff --git a/crates/oak-codec/src/proxymanager.rs b/crates/oak-codec/src/proxymanager.rs index b9c0a3952..25ab36b63 100644 --- a/crates/oak-codec/src/proxymanager.rs +++ b/crates/oak-codec/src/proxymanager.rs @@ -769,3 +769,4 @@ mod tests_extra { ); } } + ); diff --git a/crates/oak-core/Cargo.toml b/crates/oak-core/Cargo.toml index 9878ab249..eb89377f6 100644 --- a/crates/oak-core/Cargo.toml +++ b/crates/oak-core/Cargo.toml @@ -15,6 +15,11 @@ ocio-rs = { version = "0.2", features = ["bundled"] } # generation (29, matching the gpui_wgpu device the app presents with, so # the render thread's textures are directly sampleable by the UI). wgpu = "29" +# wgpu-types: the HAL descriptor types (`TextureUses`) that wgpu 29 does +# re-export through `wgpu::hal`'s descriptors but not by name (M5 external +# image import builds a `wgpu_hal::TextureDescriptor` by hand). Pinned to +# the same generation as wgpu 29. +wgpu-types = "29" log = "0.4.34" # half: f16↔f32 conversion for format-aware texture downloads (the M2 # present target is Rgba16Float; the rest of the pipeline is Rgba32Float). @@ -26,3 +31,35 @@ quick-xml = "0.41.0" # Pure-Rust image I/O for oiioutils.rs (TIFF only — the only format current # callers need). image = { version = "0.25", default-features = false, features = ["tiff"] } + +# --- M5 zero-copy hardware-decode import (platform natives) --------------- +# Linux: raw Vulkan through wgpu-hal's own `ash` generation (the version is +# pinned by wgpu-hal 29's dependency, so the `vk` types unify) + `dup(2)` +# for the DMA-BUF fds handed to `VkImportMemoryFdInfoKHR`. +[target.'cfg(target_os = "linux")'.dependencies] +libc = "0.2" + +# The raw Vulkan import path is shared by Linux (DMA-BUF) and Windows +# (D3D11 shared handle). +[target.'cfg(any(target_os = "linux", target_os = "windows"))'.dependencies] +ash = "0.38" + +# Windows: D3D11VA surfaces arrive as an `ID3D11Texture2D`; the shared NT +# handle is handed to Vulkan's `VK_EXTERNAL_MEMORY_HANDLE_TYPE_D3D11_TEXTURE` +# path. Only the raw ash structs are needed inside oak-core (the COM side +# lives in oak-codec), so no `windows` dependency here. +# +# macOS: VideoToolbox surfaces are IOSurfaces; the Metal texture wrapping +# them is built with the same objc2 generation wgpu-hal 29 uses. +[target.'cfg(target_os = "macos")'.dependencies] +objc2 = "0.6" +objc2-foundation = "0.3" +objc2-metal = { version = "0.3", features = [ + "MTLAllocation", + "MTLDevice", + "MTLPixelFormat", + "MTLResource", + "MTLTexture", + "objc2-io-surface", +] } +objc2-io-surface = "0.3" diff --git a/crates/oak-core/src/backend.rs b/crates/oak-core/src/backend.rs index 685a05867..2f32ff5a6 100644 --- a/crates/oak-core/src/backend.rs +++ b/crates/oak-core/src/backend.rs @@ -44,6 +44,12 @@ use crate::error::{Error, Result}; use crate::frame::VideoParamsPod; use crate::texture::{Frame, Texture}; +pub mod external; + +#[cfg(target_os = "linux")] +pub use external::DmaBufPlane; +pub use external::ImportGuard; + /// Backend selection preference (mapped onto wgpu backends). /// /// The choice is user-visible: the settings panel exposes a renderer @@ -222,6 +228,15 @@ struct GpuTexture { width: u32, height: u32, format: wgpu::TextureFormat, + /// The view aspect to use for sampling. Plane-imported textures + /// (Windows NV12: one texture, two plane views) register one token + /// per plane with the matching aspect; everything else is `All`. + aspect: wgpu::TextureAspect, + /// Selected array layer range for array-imported textures (Windows + /// D3D11VA frames are slices of one array texture): `Some((base, + /// count))` when this token aliases one slice. `None` = the whole + /// (single-layer) texture. + layer: Option<(u32, u32)>, } fn lock(m: &Mutex) -> MutexGuard<'_, T> { @@ -295,6 +310,9 @@ pub struct GpuContext { present: Mutex>, /// The compiled YUV→RGB pass (M5 decode import dependency). yuv: Mutex>, + /// The compiled planar (imported hardware NV12/P010) YUV→RGB pass + /// (M5): luma + interleaved chroma plane bindings. + yuv_planar: Mutex>, /// Caller-keyed 3D LUT textures (per-node color transforms; M2). color_luts: Mutex>, /// Set once this context has created a GPU resource (texture or @@ -337,6 +355,12 @@ pub fn reset_gpu_transfer_counters() { /// The process-wide shared-context slot (`shared` / `install_shared`). struct SharedSlot { decided: bool, + /// The HOST (app/UI) explicitly installed the context, as opposed to + /// the engine lazily creating one from the user config. External + /// hardware-surface imports (M5) only run on a host-installed device: + /// a lazily created private context in a worker/CLI/test must not + /// silently switch decode onto the GPU. + host_installed: bool, ctx: Option>, } @@ -345,29 +369,46 @@ fn shared_slot() -> &'static Mutex { SLOT.get_or_init(|| { Mutex::new(SharedSlot { decided: false, + host_installed: false, ctx: None, }) }) } +/// Parse an `OAK_REQUIRE_GPU` value: unset means "skipping is allowed"; +/// only `0`/`false` (case-insensitive) disable the requirement. Kept pure +/// so the policy test needs no process-wide environment mutation (which +/// would race the GPU acceptance tests reading the variable in parallel). +fn require_gpu_from_value(value: Option<&str>) -> bool { + match value { + Some(v) => !(v == "0" || v.eq_ignore_ascii_case("false")), + None => false, + } +} + /// Whether GPU-dependent tests must hard-fail when no adapter is /// available. CI sets `OAK_REQUIRE_GPU=1` on the software-Vulkan runner /// (lavapipe is present), so a degraded environment fails the suite /// instead of silently losing the GPU acceptance signal. pub fn require_gpu_adapter() -> bool { - match std::env::var("OAK_REQUIRE_GPU") { - Ok(v) => !(v == "0" || v.eq_ignore_ascii_case("false")), - Err(_) => false, + require_gpu_from_value(std::env::var("OAK_REQUIRE_GPU").ok().as_deref()) +} + +/// Handle a missing adapter: panic when `required`, otherwise log the +/// skip (the caller returns). Taking the flag as a parameter lets the +/// policy test drive the hard-fail arm without flipping the real +/// environment variable of a running test process. +fn skip_or_fail_gpu_with(required: bool, what: &str) { + if required { + panic!("no GPU adapter available for {what} (OAK_REQUIRE_GPU is set)"); } + eprintln!("no GPU adapter; skipping {what}"); } /// Handle a missing adapter in a GPU acceptance test: panic when /// `OAK_REQUIRE_GPU` is set, otherwise log the skip (the caller returns). pub fn skip_or_fail_gpu(what: &str) { - if require_gpu_adapter() { - panic!("no GPU adapter available for {what} (OAK_REQUIRE_GPU is set)"); - } - eprintln!("no GPU adapter; skipping {what}"); + skip_or_fail_gpu_with(require_gpu_adapter(), what); } /// A fresh [`GpuContext`] for a test, or `None` when no adapter exists @@ -438,6 +479,18 @@ impl GpuContext { if filterable { required_features |= wgpu::Features::FLOAT32_FILTERABLE; } + // M5 (Windows): the D3D11VA import wraps a multi-planar + // NV12/P010 texture; those formats need their features + // enabled at device creation (no-op on adapters without + // them — the import then falls back per frame). + for feature in [ + wgpu::Features::TEXTURE_FORMAT_NV12, + wgpu::Features::TEXTURE_FORMAT_P010, + ] { + if adapter.features().contains(feature) { + required_features |= feature; + } + } let (device, queue) = match pollster_block_on(adapter.request_device(&wgpu::DeviceDescriptor { label: Some("oakrender"), @@ -469,6 +522,7 @@ impl GpuContext { display_lut: Mutex::new(None), present: Mutex::new(Vec::new()), yuv: Mutex::new(None), + yuv_planar: Mutex::new(None), color_luts: Mutex::new(Vec::new()), used: AtomicBool::new(false), })); @@ -508,6 +562,7 @@ impl GpuContext { display_lut: Mutex::new(None), present: Mutex::new(Vec::new()), yuv: Mutex::new(None), + yuv_planar: Mutex::new(None), color_luts: Mutex::new(Vec::new()), used: AtomicBool::new(false), }) @@ -600,6 +655,8 @@ impl GpuContext { width: width as u32, height: height as u32, format, + aspect: wgpu::TextureAspect::All, + layer: None, }, ); Ok(token) @@ -1245,6 +1302,202 @@ impl GpuContext { Ok(program) } + /// Run the planar YUV→RGB pass (M5 zero-copy decode): the imported + /// luma plane (R8/R16, bind 0) and interleaved chroma plane + /// (R8G8/R16G16, bind 1) → one `Rgba32Float` destination. The + /// matrix/range live in `transform`, exactly like + /// [`GpuContext::run_yuv_to_rgb`]. + pub fn run_planar_yuv_to_rgb( + &self, + y: u64, + uv: u64, + dst: u64, + transform: &YuvTransform, + ) -> Result<()> { + let y_view = self.texture_view(y)?; + let uv_view = self.texture_view(uv)?; + let dst_tex = lock(&self.textures) + .get(&dst) + .cloned() + .ok_or(Error::NotFound)?; + let (y_size, uv_size) = { + let textures = lock(&self.textures); + let y = textures.get(&y).ok_or(Error::NotFound)?; + let uv = textures.get(&uv).ok_or(Error::NotFound)?; + ((y.width, y.height), (uv.width, uv.height)) + }; + let pipeline = self.planar_yuv_pipeline()?; + let m = transform.matrix; + let b = transform.offset; + let params: [f32; 16] = [ + m[0][0], + m[0][1], + m[0][2], + b[0], + m[1][0], + m[1][1], + m[1][2], + b[1], + m[2][0], + m[2][1], + m[2][2], + b[2], + uv_size.0 as f32 / y_size.0.max(1) as f32, + uv_size.1 as f32 / y_size.1.max(1) as f32, + 0.0, + 0.0, + ]; + let uniform = self.device.create_buffer(&wgpu::BufferDescriptor { + label: Some("oakrender-yuv-planar-params"), + size: 64, + usage: wgpu::BufferUsages::UNIFORM | wgpu::BufferUsages::COPY_DST, + mapped_at_creation: false, + }); + self.queue + .write_buffer(&uniform, 0, &f32_uniform_bytes(¶ms)); + let dst_view = dst_tex + .texture + .create_view(&wgpu::TextureViewDescriptor::default()); + let bind_group = self.device.create_bind_group(&wgpu::BindGroupDescriptor { + label: Some("oakrender-yuv-planar-bg"), + layout: &pipeline.layout, + entries: &[ + wgpu::BindGroupEntry { + binding: 0, + resource: wgpu::BindingResource::TextureView(&y_view), + }, + wgpu::BindGroupEntry { + binding: 1, + resource: wgpu::BindingResource::TextureView(&uv_view), + }, + wgpu::BindGroupEntry { + binding: 2, + resource: uniform.as_entire_binding(), + }, + ], + }); + let mut encoder = self + .device + .create_command_encoder(&wgpu::CommandEncoderDescriptor { + label: Some("oakrender-yuv-planar"), + }); + { + let mut pass = encoder.begin_render_pass(&wgpu::RenderPassDescriptor { + label: Some("oakrender-yuv-planar-pass"), + color_attachments: &[Some(wgpu::RenderPassColorAttachment { + view: &dst_view, + depth_slice: None, + resolve_target: None, + ops: wgpu::Operations { + load: wgpu::LoadOp::Clear(wgpu::Color::TRANSPARENT), + store: wgpu::StoreOp::Store, + }, + })], + depth_stencil_attachment: None, + timestamp_writes: None, + occlusion_query_set: None, + multiview_mask: None, + }); + pass.set_pipeline(&pipeline.pipeline); + pass.set_bind_group(0, &bind_group, &[]); + pass.draw(0..3, 0..1); + } + self.queue.submit(Some(encoder.finish())); + Ok(()) + } + + /// The planar YUV→RGB pass pipeline (built once). + fn planar_yuv_pipeline(&self) -> Result { + let mut cache = lock(&self.yuv_planar); + if let Some(p) = cache.as_ref() { + return Ok(p.clone()); + } + let plane = |binding: u32| wgpu::BindGroupLayoutEntry { + binding, + visibility: wgpu::ShaderStages::FRAGMENT, + ty: wgpu::BindingType::Texture { + sample_type: wgpu::TextureSampleType::Float { filterable: false }, + view_dimension: wgpu::TextureViewDimension::D2, + multisampled: false, + }, + count: None, + }; + let layout = self + .device + .create_bind_group_layout(&wgpu::BindGroupLayoutDescriptor { + label: Some("oakrender-yuv-planar-layout"), + entries: &[ + plane(0), + plane(1), + wgpu::BindGroupLayoutEntry { + binding: 2, + visibility: wgpu::ShaderStages::FRAGMENT, + ty: wgpu::BindingType::Buffer { + ty: wgpu::BufferBindingType::Uniform, + has_dynamic_offset: false, + min_binding_size: None, + }, + count: None, + }, + ], + }); + let vs = self + .device + .create_shader_module(wgpu::ShaderModuleDescriptor { + label: Some("oakrender-yuv-planar-vs"), + source: wgpu::ShaderSource::Wgsl(std::borrow::Cow::Borrowed(EFFECT_VS_WGSL)), + }); + let fs = self + .device + .create_shader_module(wgpu::ShaderModuleDescriptor { + label: Some("oakrender-yuv-planar-fs"), + source: wgpu::ShaderSource::Wgsl(std::borrow::Cow::Borrowed(YUV_PLANAR_WGSL)), + }); + let pipeline_layout = self + .device + .create_pipeline_layout(&wgpu::PipelineLayoutDescriptor { + label: Some("oakrender-yuv-planar-pipeline-layout"), + bind_group_layouts: &[Some(&layout)], + immediate_size: 0, + }); + let scope = self.device.push_error_scope(wgpu::ErrorFilter::Validation); + let pipeline = self + .device + .create_render_pipeline(&wgpu::RenderPipelineDescriptor { + label: Some("oakrender-yuv-planar"), + layout: Some(&pipeline_layout), + vertex: wgpu::VertexState { + module: &vs, + entry_point: Some("vs_main"), + compilation_options: Default::default(), + buffers: &[], + }, + primitive: wgpu::PrimitiveState::default(), + depth_stencil: None, + multisample: wgpu::MultisampleState::default(), + fragment: Some(wgpu::FragmentState { + module: &fs, + entry_point: Some("main"), + compilation_options: Default::default(), + targets: &[Some(wgpu::ColorTargetState { + format: wgpu::TextureFormat::Rgba32Float, + blend: None, + write_mask: wgpu::ColorWrites::ALL, + })], + }), + multiview_mask: None, + cache: None, + }); + if let Some(err) = pollster_block_on(scope.pop()) { + return Err(Error::Failed(format!( + "planar YUV pipeline validation failed: {err}" + ))); + } + let program = PresentPipeline { pipeline, layout }; + *cache = Some(program.clone()); + Ok(program) + } + /// Clear a texture to transparent black on the GPU (no CPU transfer): /// the graph compositor's starting accumulator and single-sided /// transition sides. @@ -1605,11 +1858,40 @@ impl GpuContext { return false; } } + slot.host_installed = ctx.is_some(); slot.ctx = ctx; slot.decided = true; true } + /// True when the host installed the shared device (the app's UI + /// context). The M5 hardware-surface import requires this: a lazily + /// created engine context belongs to a worker/CLI/test and must not + /// pull decode onto the GPU. + pub fn host_gpu_installed() -> bool { + lock(shared_slot()).host_installed + } + + /// Declare the process's shared context as the HOST's render device + /// (M5), creating it on first use if the engine has not yet. + /// + /// On Linux/FreeBSD the app adopts the window's wgpu device via + /// [`GpuContext::install_shared`]; on macOS/Windows gpui does not + /// expose that device, so the engine-created shared context *is* the + /// app's render device and must be marked here — otherwise the decode + /// import's host gate would make the D3D11VA/VideoToolbox rows dead + /// code (M5 audit). Worker, CLI and test processes never call this, + /// which is what keeps their decode on the CPU staging path. + pub fn mark_host_context() -> bool { + let mut slot = lock(shared_slot()); + if !slot.decided { + slot.ctx = Self::create(BackendKind::from_user_config()); + slot.decided = true; + } + slot.host_installed = slot.ctx.is_some(); + slot.ctx.is_some() + } + /// True when the app installed the shared context (as opposed to the /// engine lazily creating one from user config). Tests/UI use this to /// tell "the presenter's device" from "a private device". @@ -1931,22 +2213,46 @@ pub struct YuvTransform { } impl YuvTransform { - /// Build the transform for a luma matrix and range. + /// Build the transform for a luma matrix and range. Inputs are + /// normalized 16-bit code values (code / 65535). pub fn from_matrix(matrix: crate::colormath::YuvMatrix, full_range: bool) -> Self { + Self::from_matrix_depth(matrix, full_range, 16) + } + + /// Build the transform for a luma matrix, range and source bit depth. + /// Plane inputs are normalized code values (code / (2^bits - 1)); + /// the limited-range expansion uses the depth's own code limits + /// (16..235 luma, 16..240 chroma scaled to the depth). + pub fn from_matrix_depth( + matrix: crate::colormath::YuvMatrix, + full_range: bool, + bit_depth: u32, + ) -> Self { let (kr, kb) = matrix.kr_kb(); let kg = 1.0 - kr - kb; - // CPU reference: Y spans 219<<8 for limited, chroma 224<<8, both - // divided per code value; inputs here are code/65535. - let (ay, by, ac, bc) = if full_range { - (1.0, 0.0, 1.0, 32768.0 / 65535.0) + let max = ((1u32 << bit_depth) - 1) as f32; + // Code limits for this depth. `scale` is the 8-bit-code shift the + // CPU reference uses (depth 16 = 8-bit codes << 8, so 16 stays + // bit-identical with `colormath::yuv444p16_to_rgb_f32`). + let scale = if bit_depth > 8 { + 1u32 << (bit_depth - 8) + } else { + 1 + }; + let (y_black, y_white, c_mid, c_span) = if full_range { + (0.0, max, (1u32 << (bit_depth - 1)) as f32, max) } else { ( - 65535.0 / 56064.0, - 4096.0 / 56064.0, - 65535.0 / 57344.0, - 32768.0 / 57344.0, + (16 * scale) as f32, + (235 * scale) as f32, + (128 * scale) as f32, + (224 * scale) as f32, ) }; + let ay = max / (y_white - y_black); + let by = y_black / (y_white - y_black); + let ac = max / c_span; + let bc = c_mid / c_span; let rv = 2.0 * (1.0 - kr) * ac; let bu = 2.0 * (1.0 - kb) * ac; let gu = -(2.0 * kb * (1.0 - kb) / kg) * ac; @@ -2096,6 +2402,46 @@ fn main(@builtin(position) frag: vec4) -> @location(0) vec4 { } "#; +/// The planar (hardware-import) YUV→RGB pass (M5): luma + interleaved +/// chroma plane inputs (R8/R8G8 or R16/R16G16, normalized 0..1), one +/// `Rgba32Float` output. Same matrix/range math as [`YUV_WGSL`]. +const YUV_PLANAR_WGSL: &str = r#" +struct Params { + m0: vec4, + m1: vec4, + m2: vec4, + uv_scale: vec4, +}; +@group(0) @binding(0) var y_tex: texture_2d; +@group(0) @binding(1) var uv_tex: texture_2d; +@group(0) @binding(2) var p: Params; + +@fragment +fn main(@builtin(position) frag: vec4) -> @location(0) vec4 { + let yd = textureDimensions(y_tex); + let coord = clamp( + vec2(i32(frag.x), i32(frag.y)), + vec2(0, 0), + vec2(i32(yd.x), i32(yd.y)) - vec2(1, 1), + ); + let yv = textureLoad(y_tex, coord, 0).r; + let ud = textureDimensions(uv_tex); + let uc = clamp( + vec2(vec2(coord) * p.uv_scale.xy), + vec2(0, 0), + vec2(i32(ud.x), i32(ud.y)) - vec2(1, 1), + ); + let uv = textureLoad(uv_tex, uc, 0); + let yuv = vec3(yv, uv.r, uv.g); + let rgb = vec3( + dot(p.m0.xyz, yuv) + p.m0.w, + dot(p.m1.xyz, yuv) + p.m1.w, + dot(p.m2.xyz, yuv) + p.m2.w, + ); + return vec4(rgb, 1.0); +} +"#; + /// The fixed vertex stage for effect passes: a fullscreen triangle /// emitting `ove_texcoord`-convention UVs at location 0. UV v=0 is the /// first texture data row (the upload/download row order), so effect @@ -2368,6 +2714,9 @@ impl DisplayRenderer { let frame = unsafe { frame_from_pixels_for_upload(size, pixels, linesize) }?; ctx.upload(*token, &frame) } + // Imported hardware planes are immutable views of decoder + // memory (M5); uploading into them is never valid. + Texture::Planar(_) => Err(Error::State), Texture::Cpu(frame) => { let stride = frame.linesize_bytes(); if linesize != stride { @@ -2477,7 +2826,9 @@ impl DisplayRenderer { pub fn texture_id_of(t: &Texture) -> i32 { match t { Texture::Gpu { token, .. } => *token as i32, - Texture::Cpu(_) => 0, + // Imported planar frames have no single displayable texture id; + // the footage path resolves them before presentation. + Texture::Planar(_) | Texture::Cpu(_) => 0, } } diff --git a/crates/oak-core/src/backend/external.rs b/crates/oak-core/src/backend/external.rs new file mode 100644 index 000000000..f7e31e2cc --- /dev/null +++ b/crates/oak-core/src/backend/external.rs @@ -0,0 +1,883 @@ +// Oak Video Editor - Non-Linear Video Editor +// Copyright (C) 2026 Oak Team +// +// This program is free software: you can redistribute it and/or modify +// it under the terms of the GNU General Public License as published by +// the Free Software Foundation, either version 3 of the License, or +// (at your option) any later version. +// +// This program is distributed in the hope that it will be useful, +// but WITHOUT ANY WARRANTY; without even the implied warranty of +// MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. See the +// GNU General Public License for more details. +// +// You should have received a copy of the GNU General Public License +// along with this program. If not, see . + +//! M5: importing platform hardware surfaces as wgpu textures. +//! +//! The video decoder produces hardware surfaces (VAAPI/NVDEC/D3D11VA/ +//! VideoToolbox). Instead of downloading each frame to system memory and +//! re-uploading it, the surface's platform handle is wrapped into a +//! `wgpu::Texture` on the engine's own device: +//! +//! - **Linux**: the VAAPI surface's DMA-BUF plane (fd/offset/pitch/ +//! modifier) is imported into a raw Vulkan image +//! (`VK_EXT_external_memory_dma_buf` + `VK_EXT_image_drm_format_modifier`) +//! and wrapped through `wgpu_hal::vulkan::Device::texture_from_raw`. +//! - **Windows**: the D3D11VA texture's shared NT handle is wrapped via +//! `wgpu_hal::vulkan::Device::texture_from_d3d11_shared_handle`. +//! - **macOS**: the VideoToolbox CVPixelBuffer's IOSurface is wrapped in a +//! Metal texture (`newTextureWithDescriptor:iosurface:plane:`) and +//! handed to `wgpu_hal::metal::Device::texture_from_raw`. +//! +//! Everything here is `unsafe` platform glue; the *only* consumer is the +//! decode import in `oak-codec::gpuinterop`, and any failure must be +//! treated as "fall back to the staging path for this frame" — never as a +//! hard error that poisons later frames. + +use std::sync::Arc; + +use super::*; + +/// A DMA-BUF plane of a hardware frame (Linux/Vulkan import). +/// +/// The `fd` is borrowed: the import duplicates it before handing it to +/// Vulkan, so the caller keeps ownership of the original. +#[cfg(target_os = "linux")] +#[derive(Clone, Copy, Debug)] +pub struct DmaBufPlane { + /// DMA-BUF file descriptor. + pub fd: i32, + /// Byte offset of the plane inside the buffer object. + pub offset: u64, + /// Row pitch in bytes. + pub pitch: u64, + /// DRM format modifier (`DRM_FORMAT_MOD_*`). + pub modifier: u64, +} + +/// The Metal pixel format for an IOSurface plane (macOS). +#[cfg(target_os = "macos")] +fn iosurface_pixel_format(format: wgpu::TextureFormat) -> Option { + use objc2_metal::MTLPixelFormat; + Some(match format { + wgpu::TextureFormat::R8Unorm => MTLPixelFormat::R8Unorm, + wgpu::TextureFormat::Rg8Unorm => MTLPixelFormat::RG8Unorm, + wgpu::TextureFormat::R16Unorm => MTLPixelFormat::R16Unorm, + wgpu::TextureFormat::Rg16Unorm => MTLPixelFormat::RG16Unorm, + _ => return None, + }) +} + +/// Registration parameters for one imported texture alias: dimensions, +/// array size, layout, plane aspect and the selected array slice (M5). +#[derive(Clone, Copy)] +struct ImportedTextureDesc { + width: u32, + height: u32, + array_layers: u32, + format: wgpu::TextureFormat, + aspect: wgpu::TextureAspect, + layer: Option<(u32, u32)>, +} + +/// Caller-owned resource tied to an imported texture's lifetime (M5): +/// the decoder's hardware `AVFrame` stays alive until the GPU texture is +/// destroyed, so the decoder's surface pool cannot recycle the memory +/// while the GPU still samples it. The closure runs exactly once, right +/// before the platform resources are released. +pub type ImportGuard = Box; + +/// The plane formats the import accepts. Kept small on purpose: luma and +/// interleaved chroma for 8-bit NV12 and 16-bit-container P010. +#[cfg(target_os = "linux")] +fn import_vk_format(format: wgpu::TextureFormat) -> Option { + use ash::vk::Format; + Some(match format { + wgpu::TextureFormat::R8Unorm => Format::R8_UNORM, + wgpu::TextureFormat::Rg8Unorm => Format::R8G8_UNORM, + wgpu::TextureFormat::R16Unorm => Format::R16_UNORM, + wgpu::TextureFormat::Rg16Unorm => Format::R16G16_UNORM, + _ => return None, + }) +} + +impl GpuContext { + /// Import one DMA-BUF plane as a GPU texture (Linux/Vulkan). + /// + /// Returns the registry token. `keep_alive` (the decoder's mapped + /// `AVFrame`) runs when the GPU texture is destroyed, after the GPU is + /// done with it. + #[cfg(target_os = "linux")] + pub fn import_dmabuf_plane( + &self, + width: u32, + height: u32, + format: wgpu::TextureFormat, + plane: DmaBufPlane, + keep_alive: Option, + ) -> Result { + use ash::vk; + use wgpu::hal; + + if self.kind != BackendKind::Vulkan || width == 0 || height == 0 { + return Err(Error::State); + } + let Some(vk_format) = import_vk_format(format) else { + return Err(Error::Invalid); + }; + let guard = unsafe { self.device.as_hal::() }.ok_or(Error::State)?; + let hal_device: &hal::vulkan::Device = &guard; + let raw = hal_device.raw_device().clone(); + let _ = hal_device.raw_physical_device(); + + let layouts = [vk::SubresourceLayout { + offset: plane.offset, + size: 0, + row_pitch: plane.pitch, + array_pitch: 0, + depth_pitch: 0, + }]; + let external = vk::ExternalMemoryImageCreateInfo::default() + .handle_types(vk::ExternalMemoryHandleTypeFlags::DMA_BUF_EXT); + let mut modifier = vk::ImageDrmFormatModifierExplicitCreateInfoEXT::default() + .drm_format_modifier(plane.modifier) + .plane_layouts(&layouts); + modifier.p_next = (&external as *const vk::ExternalMemoryImageCreateInfo).cast(); + + let info = vk::ImageCreateInfo::default() + .image_type(vk::ImageType::TYPE_2D) + .format(vk_format) + .extent(vk::Extent3D { + width, + height, + depth: 1, + }) + .mip_levels(1) + .array_layers(1) + .samples(vk::SampleCountFlags::TYPE_1) + .tiling(vk::ImageTiling::DRM_FORMAT_MODIFIER_EXT) + .usage(vk::ImageUsageFlags::SAMPLED) + .sharing_mode(vk::SharingMode::EXCLUSIVE) + .initial_layout(vk::ImageLayout::UNDEFINED) + .push_next(&mut modifier); + let image = unsafe { raw.create_image(&info, None) } + .map_err(|e| Error::Failed(format!("vkCreateImage(dmabuf): {e:?}")))?; + + let req = unsafe { raw.get_image_memory_requirements(image) }; + + // Duplicate the fd: Vulkan takes ownership of the imported fd on a + // successful `vkAllocateMemory` and closes it when the memory is + // freed; the decoder keeps its own. + let dup_fd = unsafe { libc::dup(plane.fd) }; + if dup_fd < 0 { + unsafe { raw.destroy_image(image, None) }; + return Err(Error::Failed("dup(dmabuf fd) failed".into())); + } + // Pick the memory type by trying the compatible ones in order — + // `vkGetPhysicalDeviceMemoryProperties` is instance-level and + // wgpu-hal keeps its raw instance private. The compatible set is + // usually tiny (often one DEVICE_LOCAL type on discrete GPUs). + let mut import = vk::ImportMemoryFdInfoKHR::default() + .handle_type(vk::ExternalMemoryHandleTypeFlags::DMA_BUF_EXT) + .fd(dup_fd); + let mut dedicated = vk::MemoryDedicatedAllocateInfo::default().image(image); + let mut last_err = None; + let mut memory = None; + for type_index in 0..32u32 { + if req.memory_type_bits & (1 << type_index) == 0 { + continue; + } + let alloc = vk::MemoryAllocateInfo::default() + .allocation_size(req.size) + .memory_type_index(type_index) + .push_next(&mut dedicated) + .push_next(&mut import); + match unsafe { raw.allocate_memory(&alloc, None) } { + Ok(m) => { + memory = Some(m); + break; + } + Err(e) => last_err = Some((type_index, e)), + } + } + let Some(memory) = memory else { + unsafe { + libc::close(dup_fd); + raw.destroy_image(image, None); + } + return Err(Error::Failed(format!( + "vkAllocateMemory(import) failed for every compatible type: {last_err:?}" + ))); + }; + if let Err(e) = unsafe { raw.bind_image_memory(image, memory, 0) } { + unsafe { + raw.free_memory(memory, None); + raw.destroy_image(image, None); + } + return Err(Error::Failed(format!("vkBindImageMemory: {e:?}"))); + } + + let hal_desc = hal::TextureDescriptor { + label: Some("oakrender-imported-dmabuf"), + size: wgpu::Extent3d { + width, + height, + depth_or_array_layers: 1, + }, + mip_level_count: 1, + sample_count: 1, + dimension: wgpu::TextureDimension::D2, + format, + usage: wgpu_types::TextureUses::RESOURCE, + memory_flags: hal::MemoryFlags::empty(), + view_formats: Vec::new(), + }; + // wgpu-hal does not destroy external images (it cannot know how); + // the drop guard owns both the image and the imported memory, and + // carries the decoder's frame-keep-alive with it. + let raw_guard = raw.clone(); + let drop_guard: hal::DropCallback = Box::new(move || unsafe { + if let Some(keep) = keep_alive { + keep(); + } + raw_guard.destroy_image(image, None); + raw_guard.free_memory(memory, None); + }); + let hal_texture = unsafe { + hal_device.texture_from_raw( + image, + &hal_desc, + Some(drop_guard), + hal::vulkan::TextureMemory::External, + ) + }; + self.register_imported_texture::( + hal_texture, + ImportedTextureDesc { + width, + height, + array_layers: 1, + format, + aspect: wgpu::TextureAspect::All, + layer: None, + }, + ) + } + + /// Import one slice of a D3D11VA texture array shared as an NT handle + /// (Windows/Vulkan): one multi-planar image (`NV12`/`P010`) with + /// `array_size` layers, registered twice (planes 0/1) for `slice`. + /// + /// D3D11VA frames cycle through the decoder's surface array + /// (`AVFrame::data[1]` is the slice index), so the imported VkImage + /// must carry the whole array and the tokens must select their slice — + /// otherwise only every N-th frame would be importable (M5 audit). + /// + /// `keep_alive` runs when the image is destroyed; the handle itself is + /// owned by the caller (Vulkan takes a resource reference, not the + /// handle). + #[cfg(target_os = "windows")] + pub fn import_d3d11_shared_texture( + &self, + shared_handle: isize, + width: u32, + height: u32, + format: wgpu::TextureFormat, + array_size: u32, + slice: u32, + keep_alive: Option, + ) -> Result<(u64, u64)> { + use ash::vk; + use wgpu::hal; + + if self.kind != BackendKind::Vulkan || width == 0 || height == 0 { + return Err(Error::State); + } + let array_size = array_size.max(1); + if slice >= array_size { + return Err(Error::Invalid); + } + let (vk_format, required_feature) = match format { + wgpu::TextureFormat::NV12 => ( + vk::Format::G8_B8R8_2PLANE_420_UNORM, + wgpu::Features::TEXTURE_FORMAT_NV12, + ), + wgpu::TextureFormat::P010 => ( + vk::Format::G10X6_B10X6R10X6_2PLANE_420_UNORM_3PACK16, + wgpu::Features::TEXTURE_FORMAT_P010, + ), + _ => return Err(Error::Invalid), + }; + // Multi-planar formats need their device feature; without it the + // hal texture would be created but every view invalid. + if !self.device.features().contains(required_feature) { + return Err(Error::State); + } + let guard = unsafe { self.device.as_hal::() }.ok_or(Error::State)?; + let hal_device: &hal::vulkan::Device = &guard; + let raw = hal_device.raw_device().clone(); + + let mut external = vk::ExternalMemoryImageCreateInfo::default() + .handle_types(vk::ExternalMemoryHandleTypeFlags::D3D11_TEXTURE); + let info = vk::ImageCreateInfo::default() + .image_type(vk::ImageType::TYPE_2D) + .format(vk_format) + .extent(vk::Extent3D { + width, + height, + depth: 1, + }) + .mip_levels(1) + .array_layers(array_size) + .samples(vk::SampleCountFlags::TYPE_1) + .tiling(vk::ImageTiling::OPTIMAL) + .usage(vk::ImageUsageFlags::SAMPLED) + .sharing_mode(vk::SharingMode::EXCLUSIVE) + .initial_layout(vk::ImageLayout::UNDEFINED) + .push_next(&mut external); + let image = unsafe { raw.create_image(&info, None) } + .map_err(|e| Error::Failed(format!("vkCreateImage(d3d11): {e:?}")))?; + + let req = unsafe { raw.get_image_memory_requirements(image) }; + // Dedicated allocation is required for imported D3D11 textures. + let mut import = vk::ImportMemoryWin32HandleInfoKHR::default() + .handle_type(vk::ExternalMemoryHandleTypeFlags::D3D11_TEXTURE) + .handle(shared_handle); + let mut dedicated = vk::MemoryDedicatedAllocateInfo::default().image(image); + let mut last_err = None; + let mut memory = None; + for type_index in 0..32u32 { + if req.memory_type_bits & (1 << type_index) == 0 { + continue; + } + let alloc = vk::MemoryAllocateInfo::default() + .allocation_size(req.size) + .memory_type_index(type_index) + .push_next(&mut dedicated) + .push_next(&mut import); + match unsafe { raw.allocate_memory(&alloc, None) } { + Ok(m) => { + memory = Some(m); + break; + } + Err(e) => last_err = Some((type_index, e)), + } + } + let Some(memory) = memory else { + unsafe { raw.destroy_image(image, None) }; + return Err(Error::Failed(format!( + "vkAllocateMemory(import d3d11) failed: {last_err:?}" + ))); + }; + if let Err(e) = unsafe { raw.bind_image_memory(image, memory, 0) } { + unsafe { + raw.free_memory(memory, None); + raw.destroy_image(image, None); + } + return Err(Error::Failed(format!("vkBindImageMemory: {e:?}"))); + } + + let hal_desc = hal::TextureDescriptor { + label: Some("oakrender-imported-d3d11"), + size: wgpu::Extent3d { + width, + height, + depth_or_array_layers: array_size, + }, + mip_level_count: 1, + sample_count: 1, + dimension: wgpu::TextureDimension::D2, + format, + usage: wgpu_types::TextureUses::RESOURCE, + memory_flags: hal::MemoryFlags::empty(), + view_formats: Vec::new(), + }; + let raw_guard = raw.clone(); + let drop_guard: hal::DropCallback = Box::new(move || unsafe { + if let Some(keep) = keep_alive { + keep(); + } + raw_guard.destroy_image(image, None); + raw_guard.free_memory(memory, None); + }); + let hal_texture = unsafe { + hal_device.texture_from_raw( + image, + &hal_desc, + Some(drop_guard), + hal::vulkan::TextureMemory::External, + ) + }; + // Two plane views of the chosen slice of the multi-planar array: + // Y at plane 0, interleaved chroma at plane 1. + let desc = ImportedTextureDesc { + width, + height, + array_layers: array_size, + format, + aspect: wgpu::TextureAspect::Plane0, + layer: Some((slice, 1)), + }; + let (y, texture) = self.register_imported_texture_shared::( + hal_texture, desc, + )?; + let uv = self.register_texture_entry( + texture, + ImportedTextureDesc { + // The chroma plane's view is half-size; the planar pass + // derives the sampling scale from the registered sizes. + width: width.div_ceil(2), + height: height.div_ceil(2), + aspect: wgpu::TextureAspect::Plane1, + ..desc + }, + ); + Ok((y, uv)) + } + + /// Import one IOSurface plane as a Metal-backed texture (macOS): the + /// VideoToolbox pixel buffer's surface stays owned by the decoder; the + /// caller keeps it alive via the planar texture's guard. + /// + /// # Safety + /// + /// `iosurface` must be a live `IOSurfaceRef` that stays valid until + /// the imported texture is destroyed (the decoder's frame reference is + /// held by the planar texture's keep-alive guard). + #[cfg(target_os = "macos")] + pub unsafe fn import_iosurface_plane( + &self, + iosurface: *mut std::ffi::c_void, + plane: u32, + width: u32, + height: u32, + format: wgpu::TextureFormat, + ) -> Result { + use objc2::rc::Retained; + use objc2::runtime::ProtocolObject; + use objc2_io_surface::IOSurfaceRef; + use objc2_metal::{ + MTLDevice as _, MTLPixelFormat, MTLStorageMode, MTLTextureDescriptor, + MTLTextureType, MTLTextureUsage, + }; + use wgpu::hal; + + if self.kind != BackendKind::Metal || width == 0 || height == 0 || iosurface.is_null() { + return Err(Error::State); + } + let Some(pixel_format) = iosurface_pixel_format(format) else { + return Err(Error::Invalid); + }; + let guard = unsafe { self.device.as_hal::() }.ok_or(Error::State)?; + let hal_device: &hal::metal::Device = &guard; + let device: &Retained> = + hal_device.raw_device(); + + let descriptor = MTLTextureDescriptor::new(); + descriptor.setTextureType(MTLTextureType::Type2D); + descriptor.setPixelFormat(pixel_format); + descriptor.setStorageMode(MTLStorageMode::Shared); + descriptor.setUsage(MTLTextureUsage::ShaderRead); + // SAFETY: plain descriptor field setters. + unsafe { + descriptor.setWidth(width as usize); + descriptor.setHeight(height as usize); + descriptor.setMipmapLevelCount(1); + } + // SAFETY: the caller guarantees a live IOSurface for the texture's + // lifetime (the planar texture holds its keep-alive guard). + let surface: &IOSurfaceRef = unsafe { &*(iosurface as *const IOSurfaceRef) }; + // SAFETY: `surface` is a live IOSurface for the texture's lifetime + // (the caller's planar keep-alive holds the pixel buffer). + let texture = device + .newTextureWithDescriptor_iosurface_plane(&descriptor, surface, plane as usize) + .ok_or_else(|| Error::Failed("newTextureWithDescriptor:iosurface: returned nil".into()))?; + + let raw_type = MTLTextureType::Type2D; + let hal_texture = unsafe { + hal::metal::Device::texture_from_raw( + texture, + format, + raw_type, + 1, + 1, + hal::CopyExtent { + width, + height, + depth: 1, + }, + ) + }; + self.register_imported_texture::( + hal_texture, + ImportedTextureDesc { + width, + height, + array_layers: 1, + format, + aspect: wgpu::TextureAspect::All, + layer: None, + }, + ) + } + + /// Register a texture created through `create_texture_from_hal` in the + /// registry (M5 external imports; also the shared tail of every + /// platform import). + fn register_imported_texture( + &self, + hal_texture: A::Texture, + desc: ImportedTextureDesc, + ) -> Result { + self.register_imported_texture_shared::(hal_texture, desc) + .map(|(token, _)| token) + } + + /// The shared registration body; returns the token plus the owning + /// `wgpu::Texture` so multi-plane/array imports (one texture, several + /// plane aspects or slices) can register aliases without recreating + /// the texture. + fn register_imported_texture_shared( + &self, + hal_texture: A::Texture, + desc: ImportedTextureDesc, + ) -> Result<(u64, Arc)> { + let texture_desc = wgpu::TextureDescriptor { + label: Some("oakrender-imported"), + size: wgpu::Extent3d { + width: desc.width, + height: desc.height, + depth_or_array_layers: desc.array_layers.max(1), + }, + mip_level_count: 1, + sample_count: 1, + dimension: wgpu::TextureDimension::D2, + format: desc.format, + usage: wgpu::TextureUsages::TEXTURE_BINDING, + view_formats: &[], + }; + let texture = unsafe { + self.device + .create_texture_from_hal::(hal_texture, &texture_desc) + }; + self.used.store(true, Ordering::Release); + let texture = Arc::new(texture); + let token = self.register_texture_entry(texture.clone(), desc); + Ok((token, texture)) + } + + /// Insert one registry entry for `texture`. + fn register_texture_entry(&self, texture: Arc, desc: ImportedTextureDesc) -> u64 { + let token = self.next_token.fetch_add(1, Ordering::Relaxed); + lock(&self.textures).insert( + token, + GpuTexture { + texture, + width: desc.width, + height: desc.height, + format: desc.format, + aspect: desc.aspect, + layer: desc.layer, + }, + ); + token + } + + /// The view for `token`, honoring plane aspects and array slices + /// (M5 imports). + pub(crate) fn texture_view(&self, token: u64) -> Result { + let entry = lock(&self.textures) + .get(&token) + .cloned() + .ok_or(Error::NotFound)?; + let (base_array_layer, array_layer_count) = match entry.layer { + Some((base, count)) => (base, Some(count)), + None => (0, None), + }; + Ok(entry.texture.create_view(&wgpu::TextureViewDescriptor { + aspect: entry.aspect, + base_array_layer, + array_layer_count, + // wgpu-core derives `D2Array` from the TEXTURE's layer count, + // not from the view range: an array-imported slice must ask + // for `D2` explicitly or the planar pass's `D2` layout + // validation rejects the bind group (M5 desktop review). + dimension: (array_layer_count == Some(1)).then_some(wgpu::TextureViewDimension::D2), + ..Default::default() + })) + } +} + +#[cfg(test)] +mod tests { + //! Unit coverage for the safe parts of the import module. The unsafe + //! platform bodies are deliberately **not** exercised here: a real + //! import needs a hardware decoder surface (a VAAPI DMA-BUF, a D3D11VA + //! shared NT handle, a VideoToolbox IOSurface), which the + //! software-Vulkan CI runners do not have, and feeding them fake + //! handles would only assert that a driver rejects garbage. + //! + //! What covers the hardware paths instead: + //! - `import_dmabuf_plane` (Linux): a VAAPI machine running + //! `oak-render/tests/footage_import_test.rs` (DMA-BUF + DRM format + //! modifier import); the M5 real-hardware acceptance is recorded in + //! `docs/zh/plans/render-pipeline-threads-m5-branch-coverage.txt`. + //! - `import_d3d11_shared_texture` (Windows): the Windows job on a host + //! with D3D11VA; a lavapipe-only runner has no shared texture to + //! import and the decode side falls back to staging. + //! - `import_iosurface_plane` (macOS): the macOS job with VideoToolbox + //! active. + //! - `register_imported_texture*`/`texture_view`: the alias registry + //! and view helper are unit-tested below on a locally created array + //! texture (no platform import); the `Plane0`/`Plane1` aspects only + //! become valid with the multi-planar formats of a real import. + + use super::*; + use crate::backend::gpu_or_skip; + use crate::error::Error; + + /// The Linux plane format map accepts exactly the luma/chroma planes + /// the planar import feeds it and rejects everything else. + #[cfg(target_os = "linux")] + #[test] + fn import_vk_format_maps_luma_and_chroma_planes() { + use ash::vk::Format; + assert_eq!( + import_vk_format(wgpu::TextureFormat::R8Unorm), + Some(Format::R8_UNORM) + ); + assert_eq!( + import_vk_format(wgpu::TextureFormat::Rg8Unorm), + Some(Format::R8G8_UNORM) + ); + assert_eq!( + import_vk_format(wgpu::TextureFormat::R16Unorm), + Some(Format::R16_UNORM) + ); + assert_eq!( + import_vk_format(wgpu::TextureFormat::Rg16Unorm), + Some(Format::R16G16_UNORM) + ); + for rejected in [ + wgpu::TextureFormat::Rgba8Unorm, + wgpu::TextureFormat::Rgba32Float, + wgpu::TextureFormat::NV12, + wgpu::TextureFormat::P010, + ] { + assert_eq!( + import_vk_format(rejected), + None, + "{rejected:?} is not a single-plane sample format" + ); + } + } + + /// The macOS plane map accepts the interleaved 8/16-bit layouts the + /// VideoToolbox import feeds and rejects everything else. + #[cfg(target_os = "macos")] + #[test] + fn iosurface_pixel_format_maps_supported_planes() { + use objc2_metal::MTLPixelFormat; + assert_eq!( + iosurface_pixel_format(wgpu::TextureFormat::R8Unorm), + Some(MTLPixelFormat::R8Unorm) + ); + assert_eq!( + iosurface_pixel_format(wgpu::TextureFormat::Rg8Unorm), + Some(MTLPixelFormat::RG8Unorm) + ); + assert_eq!( + iosurface_pixel_format(wgpu::TextureFormat::R16Unorm), + Some(MTLPixelFormat::R16Unorm) + ); + assert_eq!( + iosurface_pixel_format(wgpu::TextureFormat::Rg16Unorm), + Some(MTLPixelFormat::RG16Unorm) + ); + assert_eq!( + iosurface_pixel_format(wgpu::TextureFormat::Rgba8Unorm), + None + ); + } + + /// `import_dmabuf_plane`'s argument validation must reject zero-sized + /// imports and non-plane formats *before* it touches Vulkan, so this + /// runs without any DMA-BUF: the plane fd (here invalid on purpose) + /// is never reached. + #[cfg(target_os = "linux")] + #[test] + fn dmabuf_import_rejects_invalid_arguments_without_importing() { + let Some(ctx) = gpu_or_skip("the DMA-BUF argument validation test") else { + return; + }; + if ctx.kind() != crate::backend::BackendKind::Vulkan { + eprintln!("SKIP: the DMA-BUF validation test needs a Vulkan context"); + return; + } + let plane = DmaBufPlane { + fd: -1, + offset: 0, + pitch: 64, + modifier: 0, + }; + let err = ctx + .import_dmabuf_plane(0, 64, wgpu::TextureFormat::R8Unorm, plane, None) + .unwrap_err(); + assert!(matches!(err, Error::State), "zero width: {err:?}"); + let err = ctx + .import_dmabuf_plane(64, 0, wgpu::TextureFormat::R8Unorm, plane, None) + .unwrap_err(); + assert!(matches!(err, Error::State), "zero height: {err:?}"); + let err = ctx + .import_dmabuf_plane(64, 64, wgpu::TextureFormat::Bgra8Unorm, plane, None) + .unwrap_err(); + assert!(matches!(err, Error::Invalid), "non-plane format: {err:?}"); + } + + /// The registry/view helper every import shares: one array texture, + /// several aliases. A single-slice alias must produce the explicit `D2` + /// view the M5 desktop review fix requests (wgpu-core would otherwise + /// derive `D2Array` from the texture's layer count) **and bind into the + /// planar pass's `D2` bind-group layout** — the actual failure point of + /// that regression: `create_view` succeeds on a derived `D2Array`, the + /// bind group does not. Broader aliases keep the array dimension (and + /// are shown to be rejected by the `D2` layout, so the guard is + /// falsifiable), and unknown tokens are `NotFound`. + #[test] + fn imported_texture_aliases_create_valid_views_and_bind() { + let Some(ctx) = gpu_or_skip("the imported-texture alias test") else { + return; + }; + // The shape a D3D11VA import registers: one 4-layer D2 texture. + let texture = Arc::new(ctx.device.create_texture(&wgpu::TextureDescriptor { + label: Some("oakrender-imported-test"), + size: wgpu::Extent3d { + width: 8, + height: 8, + depth_or_array_layers: 4, + }, + mip_level_count: 1, + sample_count: 1, + dimension: wgpu::TextureDimension::D2, + format: wgpu::TextureFormat::Rgba8Unorm, + usage: wgpu::TextureUsages::TEXTURE_BINDING | wgpu::TextureUsages::COPY_DST, + view_formats: &[], + })); + let desc = ImportedTextureDesc { + width: 8, + height: 8, + array_layers: 4, + format: wgpu::TextureFormat::Rgba8Unorm, + aspect: wgpu::TextureAspect::All, + layer: Some((2, 1)), + }; + let single = ctx.register_texture_entry(texture.clone(), desc); + let single_view = ctx.texture_view(single).expect("single-slice alias view"); + + let pair = ctx.register_texture_entry( + texture.clone(), + ImportedTextureDesc { + layer: Some((0, 2)), + ..desc + }, + ); + let pair_view = ctx.texture_view(pair).expect("two-slice alias view"); + + let whole = ctx.register_texture_entry( + texture.clone(), + ImportedTextureDesc { + layer: None, + ..desc + }, + ); + ctx.texture_view(whole).expect("whole-array alias view"); + + assert_ne!(single, pair, "aliases get distinct tokens"); + assert_ne!(pair, whole, "aliases get distinct tokens"); + assert!( + matches!(ctx.texture_view(u64::MAX), Err(Error::NotFound)), + "an unregistered token must be NotFound" + ); + + // The regression guard: the planar pass creates its bind group + // with a `D2` layout. Without `dimension: Some(D2)` the + // single-slice alias is a derived `D2Array` and this bind group + // fails validation, so the test must exercise bind-group creation + // (not just `create_view`). Use the production planar layout so the + // guard follows it if the layout changes. + let pipeline = ctx + .planar_yuv_pipeline() + .expect("the planar pass pipeline builds"); + // The chroma slot must be a single-slice `D2` alias too: slice 0 of + // the same texture serves. + let chroma = ctx.register_texture_entry( + texture, + ImportedTextureDesc { + layer: Some((0, 1)), + ..desc + }, + ); + let chroma_view = ctx.texture_view(chroma).expect("chroma alias view"); + let uniform = ctx.device.create_buffer(&wgpu::BufferDescriptor { + label: Some("oakrender-imported-test-planar-params"), + size: 64, + usage: wgpu::BufferUsages::UNIFORM, + mapped_at_creation: false, + }); + let scope = ctx.device.push_error_scope(wgpu::ErrorFilter::Validation); + let _bind_group = ctx.device.create_bind_group(&wgpu::BindGroupDescriptor { + label: Some("oakrender-imported-test-planar-bg"), + layout: &pipeline.layout, + entries: &[ + wgpu::BindGroupEntry { + binding: 0, + resource: wgpu::BindingResource::TextureView(&single_view), + }, + wgpu::BindGroupEntry { + binding: 1, + resource: wgpu::BindingResource::TextureView(&chroma_view), + }, + wgpu::BindGroupEntry { + binding: 2, + resource: uniform.as_entire_binding(), + }, + ], + }); + let error = pollster_block_on(scope.pop()); + assert!( + error.is_none(), + "single-slice aliases must bind into the planar pass's D2 layout \ + (a D2Array view is rejected here): {error:?}" + ); + + // Negative control: the broader (D2Array) alias is rejected by the + // same layout, proving mismatch in the assertion above is + // observable through the error scope. + let scope = ctx.device.push_error_scope(wgpu::ErrorFilter::Validation); + let _rejected = ctx.device.create_bind_group(&wgpu::BindGroupDescriptor { + label: Some("oakrender-imported-test-planar-bg-rejected"), + layout: &pipeline.layout, + entries: &[ + wgpu::BindGroupEntry { + binding: 0, + resource: wgpu::BindingResource::TextureView(&pair_view), + }, + wgpu::BindGroupEntry { + binding: 1, + resource: wgpu::BindingResource::TextureView(&chroma_view), + }, + wgpu::BindGroupEntry { + binding: 2, + resource: uniform.as_entire_binding(), + }, + ], + }); + let error = pollster_block_on(scope.pop()); + assert!( + error.is_some(), + "a D2Array alias must be rejected by the D2 planar layout" + ); + } +} diff --git a/crates/oak-core/src/lib.rs b/crates/oak-core/src/lib.rs index 19977c3b3..03baa77f8 100644 --- a/crates/oak-core/src/lib.rs +++ b/crates/oak-core/src/lib.rs @@ -78,4 +78,9 @@ pub mod lut; pub use handle::CHandle; pub use rational::Rational; pub use samplefmt::{PixelFormat, SampleFormat}; + +/// The wgpu generation the engine and the UI share (M2/M5). Re-exported +/// so consumers of the external-import API can name texture formats +/// without another dependency on the same major version. +pub use wgpu; pub use timerange::{TimeRange, TimeRangeList}; diff --git a/crates/oak-core/src/texture.rs b/crates/oak-core/src/texture.rs index 8d584883e..304855ca6 100644 --- a/crates/oak-core/src/texture.rs +++ b/crates/oak-core/src/texture.rs @@ -20,8 +20,8 @@ use crate::PixelFormat; use crate::Rational; use std::sync::Arc; -use crate::backend::{BackendKind, GpuContextLike}; -use crate::error::Result; +use crate::backend::{BackendKind, GpuContextLike, YuvTransform}; +use crate::error::{Error, Result}; use crate::frame::VideoParamsPod; /// A CPU frame (the payload oakcodec frames bridge into, and the value @@ -199,6 +199,146 @@ impl Drop for GpuLease { } } +/// A hardware surface's chroma layout, as imported for the planar GPU +/// textures (M5). +#[derive(Clone, Copy, Debug, PartialEq, Eq)] +pub enum PlanarFormat { + /// 8-bit NV12: R8 luma + interleaved R8G8 chroma. + Nv12, + /// 10-bit P010: 16-bit left-aligned luma + interleaved 16-bit chroma. + P010, +} + +impl PlanarFormat { + /// The single-channel luma plane format. + pub fn y_format(self) -> wgpu::TextureFormat { + match self { + PlanarFormat::Nv12 => wgpu::TextureFormat::R8Unorm, + PlanarFormat::P010 => wgpu::TextureFormat::R16Unorm, + } + } + + /// The interleaved two-channel chroma plane format. + pub fn uv_format(self) -> wgpu::TextureFormat { + match self { + PlanarFormat::Nv12 => wgpu::TextureFormat::Rg8Unorm, + PlanarFormat::P010 => wgpu::TextureFormat::Rg16Unorm, + } + } + + /// Display name (diagnostics). + pub fn name(self) -> &'static str { + match self { + PlanarFormat::Nv12 => "nv12", + PlanarFormat::P010 => "p010", + } + } + + /// The source bit depth (the limited-range expansion depth for + /// [`crate::backend::YuvTransform::from_matrix_depth`]). + pub fn bit_depth(self) -> u32 { + match self { + PlanarFormat::Nv12 => 8, + PlanarFormat::P010 => 16, + } + } +} + +/// A zero-copy hardware-decode frame (M5): the decoder's YUV hardware +/// surface imported as two GPU plane textures (luma + interleaved +/// chroma), plus everything the render thread needs to turn it into a +/// working-space RGBA texture — [`GpuContext::run_planar_yuv_to_rgb`] +/// for the matrix/range and the raw source colorimetry for the +/// source→working transform. +/// +/// This is a *transient* texture: footage decode produces it and the +/// footage path resolves it to `Texture::Gpu` before any node samples +/// it. Every other consumer sees only `Gpu`/`Cpu`. +/// +/// [`GpuContext::run_planar_yuv_to_rgb`]: crate::backend::GpuContext::run_planar_yuv_to_rgb +#[derive(Clone)] +pub struct PlanarTexture { + /// Luma plane token (registry entry, `R8Unorm`/`R16Unorm`). + pub y: u64, + /// Interleaved chroma plane token (`R8G8Unorm`/`R16G16Unorm`). + pub uv: u64, + /// Plane layout (bit depth and chroma interleave). + pub format: PlanarFormat, + /// Frame width (luma resolution). + pub width: i32, + /// Frame height (luma resolution). + pub height: i32, + /// The YUV→RGB matrix/range for the planes. + pub transform: YuvTransform, + /// Source color-primaries code point (`AVCOL_PRI_*`). + pub color_primaries: i32, + /// Source transfer-characteristic code point (`AVCOL_TRC_*`). + pub color_trc: i32, + /// The context owning both plane textures. + pub ctx: Arc, + /// Plane-token ownership: dropping the last `PlanarTexture` clone + /// releases both registry tokens (and with them the imported + /// surfaces). Never read directly. + #[allow(dead_code)] + y_lease: Arc, + #[allow(dead_code)] + uv_lease: Arc, + /// Extra lifetime guard for platforms whose GPU texture types cannot + /// carry a drop callback (macOS/Metal): the decoder's pixel-buffer + /// reference is held here and released with the planar texture. + #[allow(dead_code)] + pub keep_alive: Option>, +} + +impl PlanarTexture { + /// Build a planar texture from imported plane tokens: `planes` is + /// `(luma, chroma)`, `size` the luma dimensions. + pub fn new( + ctx: Arc, + format: PlanarFormat, + size: (i32, i32), + planes: (u64, u64), + transform: YuvTransform, + color: (i32, i32), + ) -> Self { + let (width, height) = size; + let (y, uv) = planes; + let (color_primaries, color_trc) = color; + Self { + y, + uv, + format, + width, + height, + transform, + color_primaries, + color_trc, + y_lease: GpuLease::new(ctx.clone(), y), + uv_lease: GpuLease::new(ctx.clone(), uv), + keep_alive: None, + ctx, + } + } + + /// Attach a non-CPU lifetime guard (macOS: the decoder's pixel + /// buffer), released with this texture. + pub fn set_keep_alive(&mut self, guard: Arc) { + self.keep_alive = Some(guard); + } +} + +impl std::fmt::Debug for PlanarTexture { + fn fmt(&self, f: &mut std::fmt::Formatter<'_>) -> std::fmt::Result { + f.debug_struct("PlanarTexture") + .field("y", &self.y) + .field("uv", &self.uv) + .field("format", &self.format) + .field("width", &self.width) + .field("height", &self.height) + .finish_non_exhaustive() + } +} + /// A texture: either backend-resident (GPU) or a CPU-frame wrapper. /// /// GPU textures carry an `Arc` to their [`GpuContext`](crate::backend::GpuContext) @@ -226,6 +366,9 @@ pub enum Texture { /// Shared token ownership (destroyed with the last clone). lease: Arc, }, + /// Imported hardware-decode YUV frame (M5); resolved to `Gpu` by the + /// footage path before consumers run. + Planar(PlanarTexture), /// CPU-frame wrapper (uploaded lazily by the backend). Cpu(Frame), } @@ -248,6 +391,7 @@ impl std::fmt::Debug for Texture { .field("height", height) .field("format", format) .finish(), + Texture::Planar(p) => f.debug_tuple("Texture::Planar").field(p).finish(), Texture::Cpu(frame) => f.debug_tuple("Texture::Cpu").field(frame).finish(), } } @@ -284,7 +428,7 @@ impl Texture { /// True for dummy textures. pub fn is_dummy(&self) -> bool { match self { - Texture::Gpu { .. } => false, + Texture::Gpu { .. } | Texture::Planar(_) => false, Texture::Cpu(f) => f.is_dummy(), } } @@ -294,11 +438,33 @@ impl Texture { Texture::Cpu(frame) } - /// Read back into a CPU frame (downloads for GPU textures). + /// Wrap an imported hardware YUV frame (M5). + pub fn wrap_planar(planar: PlanarTexture) -> Self { + Texture::Planar(planar) + } + + /// The imported planar frame, when this is one. + pub fn as_planar(&self) -> Option<&PlanarTexture> { + match self { + Texture::Planar(p) => Some(p), + _ => None, + } + } + + /// True for imported hardware YUV frames (M5) — the footage path's + /// resolve trigger. + pub fn is_planar(&self) -> bool { + matches!(self, Texture::Planar(_)) + } + + /// Read back into a CPU frame (downloads for GPU textures). An + /// unresolved planar texture is an error: the caller must resolve + /// (YUV→RGB) through the footage path first. pub fn to_frame(&self) -> Result { match self { Texture::Cpu(f) => Ok(f.clone()), Texture::Gpu { token, ctx, .. } => ctx.download(*token), + Texture::Planar(_) => Err(Error::State), } } @@ -306,14 +472,18 @@ impl Texture { pub fn size(&self) -> (i32, i32) { match self { Texture::Gpu { width, height, .. } => (*width, *height), + Texture::Planar(p) => (p.width, p.height), Texture::Cpu(f) => (f.width, f.height), } } - /// The texture's pixel format. + /// The texture's pixel format. A planar texture reports `F32`: it is + /// transient, and the footage path always resolves it to an F32 RGBA + /// GPU texture before any consumer sees it. pub fn format(&self) -> PixelFormat { match self { Texture::Gpu { format, .. } => *format, + Texture::Planar(_) => PixelFormat::F32, Texture::Cpu(f) => f.format, } } @@ -322,6 +492,7 @@ impl Texture { pub fn backend(&self) -> BackendKind { match self { Texture::Gpu { backend, .. } => *backend, + Texture::Planar(p) => p.ctx.kind(), Texture::Cpu(_) => BackendKind::Cpu, } } diff --git a/crates/oak-plugin/src/instance.rs b/crates/oak-plugin/src/instance.rs index 69ccc9c09..6527fc898 100644 --- a/crates/oak-plugin/src/instance.rs +++ b/crates/oak-plugin/src/instance.rs @@ -1216,3 +1216,4 @@ pub struct ClipPreferences { /// field 处理模式（kOfxImageField*）。 pub field: String, } + /// field 处理模式（kOfxImageField*）。 diff --git a/crates/oak-plugin/src/render_driver.rs b/crates/oak-plugin/src/render_driver.rs index d677720c6..9d8177447 100644 --- a/crates/oak-plugin/src/render_driver.rs +++ b/crates/oak-plugin/src/render_driver.rs @@ -623,3 +623,8 @@ pub(crate) fn write_output_frame(dst: &mut Texture, image: &Image) -> crate::err } Ok(()) } + ctx.upload(*token, &frame) + .map_err(|e| Error::Failed(format!("输出纹理上传失败：{e}")))?; + } ++ // 未解析的平面纹理不会成为插件输出目标（解码路径会先解析）。 ++ Texture::Planar(_) => { diff --git a/crates/oak-plugin/src/suites/property.rs b/crates/oak-plugin/src/suites/property.rs index bf7d8a51e..b51fe85a2 100644 --- a/crates/oak-plugin/src/suites/property.rs +++ b/crates/oak-plugin/src/suites/property.rs @@ -769,3 +769,4 @@ pub fn suite_v1() -> &'static PropertySuiteV1 { get_dimension: prop_get_dimension, }) } + get_dimension: prop_get_dimension, diff --git a/crates/oak-render/src/eval.rs b/crates/oak-render/src/eval.rs index 7d96b7b4d..bf2d35a42 100644 --- a/crates/oak-render/src/eval.rs +++ b/crates/oak-render/src/eval.rs @@ -345,7 +345,9 @@ impl RenderEvalHooks { height, .. } => (*token, ctx.clone(), *width, *height), - Texture::Cpu(_) => unreachable!("handled above"), + // Imported planar frames never reach node processing (the + // footage path resolves them first); pass through defensively. + Texture::Cpu(_) | Texture::Planar(_) => return Ok(tex), }; let concrete = ctx .as_any() @@ -869,6 +871,11 @@ impl RenderEvalHooks { scratch.push(token); (token, (frame.width, frame.height)) } + // Resolved by the footage path; never a shader input. + Texture::Planar(_) => { + warn("planar texture reached a shader job input (unresolved)"); + return None; + } }; // The pass size follows the effect input's texture (C++ the // job's video params = the main input size); any other bound @@ -1125,7 +1132,7 @@ const MAX_CACHED_DECODERS: usize = 6; /// LRU tick source for [`DECODERS`]. static DECODER_TICK: std::sync::atomic::AtomicU64 = std::sync::atomic::AtomicU64::new(1); -/// Count of [`render_footage_frame_inner`] entries, i.e. of real codec work +/// Count of [`render_footage_frame_inner_opts`] entries, i.e. of real codec work /// on the footage path (frame-cache hits and decode-service LRU hits do not /// count). Visible for the decode-service tests, which use it to tell a /// cached frame from a re-decode. @@ -1151,28 +1158,76 @@ pub fn reset_decode_invocations() { /// on every pre-render restart and on repeat plays). const MAX_CACHED_FRAMES: usize = 24; -/// Decoded-frame LRU: `(filename, stream, time, w, h)` -> F32 CPU frame. -/// Frame data is the expensive part (a 1080p frame ≈ 31 MB); the decoder -/// session cache alone still re-decodes every `render_footage_frame`. -/// The size is part of the key: the same media at a different target -/// resolution is a different frame (an interleaved source-monitor/proxy -/// request must not reuse a wrongly-sized pixel buffer). -static DECODED_FRAMES: std::sync::OnceLock< - std::sync::Mutex< - std::collections::HashMap<(String, i32, (i64, i64), i32, i32), (Frame, u64)>, - >, -> = std::sync::OnceLock::new(); +/// Cap on cached PLANAR hardware frames (M5 audit): each planar entry +/// pins a decoder surface through its keep-alive guard, and the driver's +/// surface pool is finite. Planar textures only need to bridge the decode +/// thread → resolve, so they get a much smaller budget than the general +/// frame LRU. Eviction drops only the import; the decoder session cache +/// still holds the raw frame, so a re-request re-imports (no re-decode). +const MAX_CACHED_PLANAR_FRAMES: usize = 4; -fn decoded_frames() -> std::sync::MutexGuard< - 'static, - std::collections::HashMap<(String, i32, (i64, i64), i32, i32), (Frame, u64)>, -> { +/// Decoded-frame LRU key: `(filename, stream, time, w, h)`. The size is +/// part of the key: the same media at a different target resolution is a +/// different frame (an interleaved source-monitor/proxy request must not +/// reuse a wrongly-sized pixel buffer). +type DecodedFrameKey = (String, i32, (i64, i64), i32, i32); + +/// Decoded-frame LRU map. Values are decoded textures (CPU frame, or the +/// M5 imported planar texture when the hardware import took the frame) +/// plus their LRU tick. Frame data is the expensive part (a 1080p F32 +/// frame ≈ 31 MB); the decoder session cache alone still re-decodes every +/// `render_footage_frame`. +type DecodedFrameCache = std::collections::HashMap; + +static DECODED_FRAMES: std::sync::OnceLock> = + std::sync::OnceLock::new(); + +fn decoded_frames() -> std::sync::MutexGuard<'static, DecodedFrameCache> { DECODED_FRAMES .get_or_init(|| std::sync::Mutex::new(std::collections::HashMap::new())) .lock() .unwrap_or_else(|e| e.into_inner()) } +/// Insert `texture` under `key` with the general and planar caps applied. +fn insert_cached_frame(cache: &mut DecodedFrameCache, key: DecodedFrameKey, texture: Texture, tick: u64) { + if cache.contains_key(&key) { + return; + } + while cache.len() >= MAX_CACHED_FRAMES { + let Some(victim) = cache + .iter() + .filter(|(_, (_, t))| *t > 0) + .min_by_key(|(_, (_, t))| *t) + .map(|(k, _)| k.clone()) + else { + break; + }; + cache.remove(&victim); + } + if texture.is_planar() { + prune_planar_frames(cache, MAX_CACHED_PLANAR_FRAMES.saturating_sub(1)); + } + cache.insert(key, (texture, tick)); +} + +/// Evict the least-recently-used planar entries until at most `keep` +/// remain (hardware-surface residency bound, M5 audit). +fn prune_planar_frames(cache: &mut DecodedFrameCache, keep: usize) { + let mut planar: Vec<(DecodedFrameKey, u64)> = cache + .iter() + .filter(|(_, (t, _))| t.is_planar()) + .map(|(k, (_, tick))| (k.clone(), *tick)) + .collect(); + if planar.len() <= keep { + return; + } + planar.sort_by_key(|(_, tick)| *tick); + for (key, _) in planar.drain(..planar.len() - keep) { + cache.remove(&key); + } +} + /// Time-key rational (the frame-LRU's deterministic key half). fn time_key(t: &Rational) -> (i64, i64) { (t.numerator(), t.denominator()) @@ -1272,24 +1327,67 @@ pub fn render_footage_frame( format: PixelFormat, ) -> Result { let started = std::time::Instant::now(); - let result = match crate::pipeline::decode_service() { - Some(service) => { - let request = crate::pipeline::DecodeRequest { - filename: filename.to_string(), + let decode = |allow_import: bool| -> Result { + match crate::pipeline::decode_service() { + Some(service) => { + let request = crate::pipeline::DecodeRequest { + filename: filename.to_string(), + stream_index, + time, + size, + format, + allow_import: true, + }; + match service.request(request) { + Some(result) => result, + // The service went away (shutdown) mid-flight: decode here + // rather than failing the frame. + None => render_footage_frame_inner_opts( + filename, + stream_index, + time, + size, + format, + allow_import, + ), + } + } + None => render_footage_frame_inner_opts( + filename, stream_index, time, size, format, - }; - match service.request(request) { - Some(result) => result, - // The service went away (shutdown) mid-flight: decode here - // rather than failing the frame. - None => render_footage_frame_inner(filename, stream_index, time, size, format), - } + allow_import, + ), } - None => render_footage_frame_inner(filename, stream_index, time, size, format), }; + let mut result = decode(true); + // M5: an imported hardware frame is planar YUV; resolve it to + // working-space RGBA on the GPU before it leaves the footage path. + // A resolution failure is a per-frame fallback: re-decode through the + // CPU scaler with the import disabled (the decoder session cache makes + // this a transfer, not a second decode). + if let Ok(texture) = &result { + if texture.is_planar() { + result = match resolve_planar_footage(texture) { + Ok(resolved) => Ok(resolved), + Err(err) => { + eprintln!("planar footage resolve failed, staging fallback: {err:#}"); + // Bypass the decode service: its cache still holds the + // planar entry that just failed to resolve. + render_footage_frame_inner_opts( + filename, + stream_index, + time, + size, + format, + false, + ) + } + }; + } + } if std::env::var_os("OAK_PERF").is_some() { eprintln!( "[decode] {:?} s{} time {}/{} ({:.3}s) size {:?} -> {:.3}s {:?}", @@ -1306,15 +1404,84 @@ pub fn render_footage_frame( result } -/// The synchronous decode behind [`render_footage_frame`] (frame-LRU → -/// decoder session → codec → F32 frame). `pub(crate)` because the decode -/// service runs exactly this on its own thread. -pub(crate) fn render_footage_frame_inner( +/// The CPU-staging decode for consumers that composite on the CPU (the +/// montage compositor): same service/FRU semantics as +/// [`render_footage_frame`], but the M5 zero-copy import is disabled and +/// the result is always a CPU frame. +/// +/// This is not just an optimization: the montage compositor runs on the +/// CPU, so an imported `Texture::Gpu` would have to be downloaded again — +/// and before this entry point existed it was silently SKIPPED (the M5 +/// audit's all-black montage bug). A planar texture must never reach a +/// CPU consumer here. +pub(crate) fn render_footage_frame_staged( filename: &str, stream_index: i32, time: Rational, size: (i32, i32), format: PixelFormat, +) -> Result { + let result = match crate::pipeline::decode_service() { + Some(service) => { + let request = crate::pipeline::DecodeRequest { + filename: filename.to_string(), + stream_index, + time, + size, + format, + allow_import: false, + }; + match service.request(request) { + Some(result) => result, + None => render_footage_frame_inner_opts( + filename, + stream_index, + time, + size, + format, + false, + ), + } + } + None => render_footage_frame_inner_opts( + filename, + stream_index, + time, + size, + format, + false, + ), + }; + // Defensive: a planar texture (a stale service entry from an older + // key layout) can never satisfy a staging request — re-decode CPU. + match result { + Ok(texture) if texture.is_planar() => render_footage_frame_inner_opts( + filename, + stream_index, + time, + size, + format, + false, + ), + other => other, + } +} + +/// The synchronous decode behind [`render_footage_frame`] (frame-LRU → +/// decoder session → codec → F32 frame). `pub(crate)` because the decode +/// service runs exactly this on its own thread, with the request's own +/// `allow_import` flag (the montage requests stage on purpose, §M5 audit). +/// +/// `allow_import` false forces the CPU staging path: the resolve step's +/// fallback re-decodes with it off, and the montage compositor never +/// wants a GPU frame. +pub(crate) fn render_footage_frame_inner_opts( + filename: &str, + stream_index: i32, + time: Rational, + size: (i32, i32), + format: PixelFormat, + allow_import: bool, ) -> Result { // (file, stream, time) repeatedly (pre-render restarts, repeated // scale-up at the same time, graph + montage interleaving); the @@ -1327,10 +1494,16 @@ pub(crate) fn render_footage_frame_inner( { let mut cache = decoded_frames(); let tick = DECODER_TICK.fetch_add(1, std::sync::atomic::Ordering::Relaxed); - let frame = cache.get(&cache_key).map(|(f, _)| f.clone()); - if let Some(frame) = frame { - cache.insert(cache_key.clone(), (frame.clone(), tick)); - return Ok(Texture::wrap_frame(frame)); + let texture = cache.get(&cache_key).map(|(f, _)| f.clone()); + if let Some(texture) = texture { + // A planar entry cannot satisfy a staging request (the resolve + // step failed and asked for CPU pixels): treat it as a miss and + // decode through the scaler instead. + if allow_import || !texture.is_planar() { + cache.insert(cache_key.clone(), (texture.clone(), tick)); + return Ok(texture); + } + cache.remove(&cache_key); } } // Past the frame cache: this call really goes to the codec (the @@ -1355,6 +1528,20 @@ pub(crate) fn render_footage_frame_inner( None }, }; + + // M5: hardware frames are offered to the zero-copy import first; the + // CPU retrieval below is the per-frame staging fallback (the decoder + // session cache holds the raw surface, so the fallback never + // re-decodes). + if allow_import { + if let Ok(Some(imported)) = decoder.retrieve_video_frame_gpu(¶ms) { + let mut cache = decoded_frames(); + let tick = DECODER_TICK.fetch_add(1, std::sync::atomic::Ordering::Relaxed); + insert_cached_frame(&mut cache, cache_key, imported.clone(), tick); + return Ok(imported); + } + } + let decoded = decoder .retrieve_video_frame(¶ms) .map_err(|e| Error::Failed(format!("footage decode at {time:?}: {e:?}")))?; @@ -1398,24 +1585,127 @@ pub(crate) fn render_footage_frame_inner( // by default; the legacy sRGB working space keeps the pass-through). convert_decoded_to_working(&mut dst, &decoded); // Memoize the finished working-space frame (LRU-capped). + let texture = Texture::wrap_frame(dst); { let mut cache = decoded_frames(); let tick = DECODER_TICK.fetch_add(1, std::sync::atomic::Ordering::Relaxed); - if !cache.contains_key(&cache_key) { - if cache.len() >= MAX_CACHED_FRAMES { - if let Some(victim) = cache - .iter() - .filter(|(_, (_, t))| *t > 0) - .min_by_key(|(_, (_, t))| *t) - .map(|(k, _)| k.clone()) - { - cache.remove(&victim); - } + insert_cached_frame(&mut cache, cache_key, texture.clone(), tick); + } + Ok(texture) +} + +/// Resolve an imported planar hardware frame (M5) to working-space RGBA +/// on the GPU: the YUV→RGB pass (matrix + range from the frame's own +/// colorimetry) followed by the source→working transform. In the legacy +/// sRGB working space the transform is a pass-through (matching the CPU +/// path's `convert_decoded_to_working`); otherwise it is applied as a CPU +/// reference-baked 3D LUT, exactly like the OCIO color transforms. +fn resolve_planar_footage(texture: &Texture) -> Result { + let planar = texture + .as_planar() + .ok_or_else(|| Error::Failed("resolve_planar_footage: not planar".into()))?; + let Some(gpu) = planar + .ctx + .as_any() + .and_then(|a| a.downcast_ref::()) + else { + return Err(Error::Failed( + "planar texture belongs to a non-wgpu context".into(), + )); + }; + let (w, h) = (planar.width.max(1), planar.height.max(1)); + let dst = gpu + .create_texture(w, h) + .map_err(|e| Error::Failed(format!("planar resolve target: {e:?}")))?; + if let Err(e) = gpu.run_planar_yuv_to_rgb(planar.y, planar.uv, dst, &planar.transform) { + gpu.destroy_texture(dst); + return Err(Error::Failed(format!("planar YUV pass: {e:?}"))); + } + let token = match footage_working_lut(planar.color_primaries, planar.color_trc) { + Some((key, lut)) => match gpu.apply_color_lut(dst, &key, &lut) { + Ok(transformed) => { + gpu.destroy_texture(dst); + transformed + } + Err(e) => { + eprintln!("footage working-space LUT failed, using source RGB: {e:?}"); + dst + } + }, + None => dst, + }; + Ok(Texture::gpu(planar.ctx.clone(), token, w, h, PixelFormat::F32)) +} + +/// The source→working-space 3D LUT for a decoded frame (M5 GPU import +/// path): baked from the same CPU reference (`colormath::decode_to_acescg`) +/// the staging path uses, cached per colorimetry. `None` in the legacy +/// sRGB working space (pass-through) or when the LUT cannot be built. +fn footage_working_lut( + color_primaries: i32, + color_trc: i32, +) -> Option<(String, std::sync::Arc)> { + use oak_core::colormath::WorkingColorSpace; + if oak_core::color::pipeline_working_space() == WorkingColorSpace::SrgbLegacy { + return None; + } + let key = format!("footage/{color_primaries}/{color_trc}"); + type Cache = std::sync::Mutex< + std::collections::HashMap>, + >; + static CACHE: std::sync::OnceLock = std::sync::OnceLock::new(); + let mut cache = CACHE + .get_or_init(|| std::sync::Mutex::new(std::collections::HashMap::new())) + .lock() + .unwrap_or_else(|e| e.into_inner()); + if let Some(lut) = cache.get(&key) { + return Some((key, lut.clone())); + } + let lut = build_footage_working_lut(color_primaries, color_trc)?; + if cache.len() >= 16 { + cache.clear(); + } + cache.insert(key.clone(), lut.clone()); + Some((key, lut)) +} + +/// Bake the source→ACEScg transform into a 3D LUT over the display +/// domain (the same domain the color-transform LUTs use). +fn build_footage_working_lut( + color_primaries: i32, + color_trc: i32, +) -> Option> { + use oak_core::colormath::{decode_to_acescg, source_primaries_from_av, source_transfer_from_av}; + use oak_core::lut::Lut3d; + let edge = Lut3d::DISPLAY_EDGE; + let (lo, hi) = (Lut3d::DISPLAY_LO, Lut3d::DISPLAY_HI); + let n = (edge as usize).pow(3); + let mut samples = vec![0.0f32; n * 4]; + let step = |i: usize, axis: usize| -> f32 { + let t = i as f32 / (edge - 1) as f32; + lo[axis] + (hi[axis] - lo[axis]) * t + }; + for b in 0..edge as usize { + for g in 0..edge as usize { + for r in 0..edge as usize { + let idx = ((b * edge as usize + g) * edge as usize + r) * 4; + samples[idx] = step(r, 0); + samples[idx + 1] = step(g, 1); + samples[idx + 2] = step(b, 2); + samples[idx + 3] = 1.0; } - cache.insert(cache_key.clone(), (dst.clone(), tick)); } } - Ok(Texture::wrap_frame(dst)) + decode_to_acescg( + &mut samples, + source_primaries_from_av(color_primaries), + source_transfer_from_av(color_trc), + ); + let mut data = Vec::with_capacity(n * 3); + for px in samples.chunks_exact(4) { + data.extend_from_slice(&px[..3]); + } + Some(std::sync::Arc::new(Lut3d { edge, lo, hi, data })) } /// Convert a decoded footage frame (display-referred RGB in the source's @@ -1532,6 +1822,11 @@ fn composite_tracks_gpu( scratch.push(t); t } + Texture::Planar(_) => { + return Err(Error::Failed( + "unresolved planar texture in composite".into(), + )) + } }; let out = ctx.create_texture(w, h)?; ctx.run_shader_pass(&program, &[], &[current, src_token], out)?; @@ -2503,7 +2798,8 @@ pub fn render_montage_frame_into( continue; } let media_time = clip.media_in + (time - clip.in_time); - let decoded = render_footage_frame( + // CPU compositor: stage on purpose (never an imported GPU frame). + let decoded = render_footage_frame_staged( &clip.filename, clip.stream_index, media_time, @@ -2514,9 +2810,24 @@ pub fn render_montage_frame_into( // (C++ semantics: the clip texture passes through the chain // bottom-up, the chain top feeds the track composite). let effected = apply_clip_effects(decoded, clip, time); + // The compositor is CPU-side. The decode is staged above, but a + // clip effect may still hand back a GPU texture: read it back + // (an explicit CPU boundary) instead of silently dropping the + // clip (M5 audit: that skip produced all-black sequences). + let downloaded: Frame; let (src_data, src_stride) = match &effected { Texture::Cpu(src) => (&src.data, src.linesize_bytes() as i32), - _ => continue, + Texture::Gpu { .. } => { + downloaded = effected.to_frame().map_err(|e| { + Error::Failed(format!("montage GPU readback failed: {e:?}")) + })?; + (&downloaded.data, downloaded.linesize_bytes() as i32) + } + Texture::Planar(_) => { + return Err(Error::Failed( + "unresolved planar texture reached the montage compositor".into(), + )) + } }; composite_over(dst, dst_stride, w, h, src_data, src_stride, clip.gain); } @@ -4420,3 +4731,4 @@ mod tests { } ++ static LOCK: Mutex<()> = Mutex::new(()); diff --git a/crates/oak-render/src/pipeline.rs b/crates/oak-render/src/pipeline.rs index 5fbbccdab..61a3f4d09 100644 --- a/crates/oak-render/src/pipeline.rs +++ b/crates/oak-render/src/pipeline.rs @@ -153,6 +153,9 @@ fn playback_decode_requests( time: clip.media_in + (time - clip.in_time), size, format: PixelFormat::F32, + // The montage compositor is CPU-side: importing would only + // be downloaded again. + allow_import: false, }); } if let Some((filename, stream_index)) = ¶ms.footage { @@ -165,6 +168,7 @@ fn playback_decode_requests( time, size, format: params.force_format.unwrap_or(PixelFormat::F32), + allow_import: true, }); } requests @@ -191,6 +195,11 @@ pub struct DecodeRequest { pub size: (i32, i32), /// Target pixel format. pub format: PixelFormat, + /// Whether the M5 zero-copy hardware import may serve this request. + /// The CPU montage compositor sets `false` (it needs CPU pixels); the + /// flag is part of the cache key, so the service never hands an + /// imported texture to a staging consumer (M5 audit). + pub allow_import: bool, } /// Decode-service counters (M1's evidence that the service really decodes @@ -553,12 +562,13 @@ fn prefetch_into( /// decode thread. fn decode(request: &DecodeRequest, inner: &DecodeInner) -> Result { inner.counters.decodes.fetch_add(1, Ordering::Relaxed); - let result = crate::eval::render_footage_frame_inner( + let result = crate::eval::render_footage_frame_inner_opts( &request.filename, request.stream_index, request.time, request.size, request.format, + request.allow_import, ); if result.is_err() { inner.counters.errors.fetch_add(1, Ordering::Relaxed); diff --git a/crates/oak-task/src/export.rs b/crates/oak-task/src/export.rs index 4185b0616..9495cad2d 100644 --- a/crates/oak-task/src/export.rs +++ b/crates/oak-task/src/export.rs @@ -295,6 +295,13 @@ impl ExportTask { "Render frame readback for the encoder failed: {e:?}" )) })?, + // Unresolved planar frames never reach an encoder; the + // decode path resolves them before delivery. + Texture::Planar(_) => { + return Err(Error::Failed( + "unresolved planar texture for the encoder".into(), + )) + } }; let frame = &frame; let params = CommonVideoParams::new_basic( diff --git a/crates/oak-task/src/nodeops.rs b/crates/oak-task/src/nodeops.rs index 05f23cfdc..c7c6eb45a 100644 --- a/crates/oak-task/src/nodeops.rs +++ b/crates/oak-task/src/nodeops.rs @@ -1181,3 +1181,4 @@ pub fn add_track_command(project: ProjectRef, list: NodeId) -> UndoCommand { kind: TrackType::Video, }) } + kind: TrackType::Video, diff --git a/crates/oak-worker/src/worker.rs b/crates/oak-worker/src/worker.rs index 6e2641f78..5b0319853 100644 --- a/crates/oak-worker/src/worker.rs +++ b/crates/oak-worker/src/worker.rs @@ -1308,6 +1308,16 @@ fn render_f32_into( } } } + Ok(oak_core::texture::Texture::Planar(_)) => { + // The footage path resolves imported planar + // frames before returning; a planar texture + // here is a bug, not a frame. + warn_graph_fallback( + spec.viewer_node, + "unresolved planar texture", + ); + None + } Err(e) => { warn_graph_fallback(spec.viewer_node, &e.to_string()); None @@ -1356,6 +1366,11 @@ fn render_f32_into( gpu @ oak_core::texture::Texture::Gpu { .. } => gpu .to_frame() .map_err(|e| format!("decode readback: {e}"))?, + // Resolved by the footage path; a planar frame cannot cross + // into the shm slot without a readback. + oak_core::texture::Texture::Planar(_) => { + return Err("unresolved planar decode texture".to_string()) + } }; let src_stride = frame.linesize_bytes() as usize; let row_bytes = (w as usize) * 16;