xgpui_gpu_interop/
lib.rs

1//! Experimental zero-copy GPU interop for xgpui: show frames from a GPU engine
2//! you own in xgpui windows, the way an engine presents to a swapchain. See the
3//! [GPU interop guide](https://xgpui.dev/gpu-interop.html).
4//!
5//! The engine keeps its own device, queue and thread. An [`EngineSurface`] owns
6//! a fixed ring of shareable slots:
7//!
8//! 1. `acquire()` a writable slot, or get [`AcquireError::WouldBlock`] when
9//!    every slot is still in use (bounded backpressure).
10//! 2. Render into the slot on the engine's GPU and submit a signal of a sync
11//!    point after that work.
12//! 3. `present(slot, sync)` publishes the slot. It is never written again
13//!    until every retained scene that sampled it is gone and the compositor's
14//!    GPU sampling has completed; then it returns to the ring.
15//!
16//! The host composites the latest presented slot as ordinary custom content,
17//! after a GPU-side wait on the sync point. Nothing is copied through the CPU.
18//!
19//! Contract: call `present` only after the engine has *submitted* the work that
20//! signals the sync point. A sync point whose signal is never submitted can
21//! stall composition.
22//!
23//! Implemented: macOS (IOSurface slots, `MTLSharedEvent` sync); with feature
24//! `wgpu`, Linux (opaque-fd Vulkan memory, timeline semaphore) and Windows
25//! (shared NT-handle D3D12 textures and fence, opened by the D3D11
26//! compositor). The compositor waits for each frame on the GPU.
27//!
28//! The engine must run on the compositor's GPU; otherwise the surface reports
29//! [`Unsupported`]. On a machine with two GPUs, nothing yet helps an engine pick
30//! the right one (issue #253).
31use gpui::{DevicePixels, Size, size};
32
33pub(crate) mod heavy;
34// Only the surface uses the ring; Linux and Windows have none without `wgpu`.
35#[cfg_attr(
36    not(any(
37        target_os = "macos",
38        all(any(target_os = "linux", target_os = "windows"), feature = "wgpu")
39    )),
40    allow(dead_code)
41)]
42mod ring;
43pub use heavy::{HeavyEngine, HeavyStatus};
44
45#[cfg(target_os = "macos")]
46mod macos;
47#[cfg(target_os = "macos")]
48use macos as platform;
49#[cfg(target_os = "macos")]
50pub use macos::*;
51
52#[cfg(all(target_os = "linux", feature = "wgpu"))]
53mod linux;
54#[cfg(all(target_os = "linux", feature = "wgpu"))]
55use linux as platform;
56#[cfg(all(target_os = "linux", feature = "wgpu"))]
57pub use linux::*;
58
59#[cfg(all(target_os = "windows", feature = "wgpu"))]
60mod d3d;
61#[cfg(all(target_os = "windows", feature = "wgpu"))]
62use d3d as platform;
63#[cfg(all(target_os = "windows", feature = "wgpu"))]
64pub use d3d::*;
65
66#[cfg(any(
67    target_os = "macos",
68    all(any(target_os = "linux", target_os = "windows"), feature = "wgpu")
69))]
70mod surface;
71#[cfg(any(
72    target_os = "macos",
73    all(any(target_os = "linux", target_os = "windows"), feature = "wgpu")
74))]
75pub use surface::{EngineSurface, SurfaceSlot};
76
77/// The wgpu version engine adapters are built against; use it for engine devices.
78#[cfg(all(any(target_os = "macos", target_os = "windows"), feature = "wgpu"))]
79pub use ::wgpu;
80/// The wgpu version engine adapters are built against; use it for engine devices.
81#[cfg(all(target_os = "linux", feature = "wgpu"))]
82pub use gpui_wgpu::wgpu;
83
84/// Pixel storage and color interpretation of every slot in a surface.
85#[derive(Clone, Copy, Debug, PartialEq, Eq)]
86pub enum SurfaceFormat {
87    /// BGRA8, straight alpha, sRGB-encoded channels. Works in every window.
88    Bgra8Srgb,
89    /// RGBA16 float, associated alpha, linear Display P3/D65. Shown in wide
90    /// gamut in a window that negotiated `WindowColorMode::WideGamutSdr`
91    /// (macOS and Windows), and as a mapped sRGB preview in any other window
92    /// and on Linux (#146). Slots also allow storage (compute) writes.
93    Rgba16FloatLinearP3,
94}
95impl SurfaceFormat {
96    /// Bytes per pixel.
97    pub fn bytes_per_pixel(self) -> u32 {
98        match self {
99            Self::Bgra8Srgb => 4,
100            Self::Rgba16FloatLinearP3 => 8,
101        }
102    }
103    /// The color interpretation the host applies when compositing into a
104    /// window whose effective color mode is (`wide`) or is not wide-gamut SDR.
105    pub fn color_mode(self, wide: bool) -> gpui::ImageColorMode {
106        match self {
107            Self::Bgra8Srgb => gpui::ImageColorMode::EncodedSrgb,
108            #[cfg(any(target_os = "macos", target_os = "windows"))]
109            Self::Rgba16FloatLinearP3 if wide => gpui::ImageColorMode::LinearP3Wide,
110            Self::Rgba16FloatLinearP3 => {
111                let _ = wide;
112                gpui::ImageColorMode::LinearP3Preview
113            }
114        }
115    }
116    /// Whether engines may bind slots as storage textures (compute writes).
117    pub fn supports_storage(self) -> bool {
118        matches!(self, Self::Rgba16FloatLinearP3)
119    }
120}
121
122/// Fixed properties of a surface's slots.
123#[derive(Clone, Copy, Debug, PartialEq, Eq)]
124pub struct SurfaceDesc {
125    pub width: u32,
126    pub height: u32,
127    pub format: SurfaceFormat,
128    /// Ring length: 2..=8. Use at least 3 for continuous animation, plus
129    /// one per additional window showing the surface: a window that is not
130    /// drawing (for example minimized) keeps the frame it last showed in its
131    /// retained scene, which pins that slot until the window draws again.
132    pub slots: usize,
133}
134impl SurfaceDesc {
135    /// Validate dimensions, ring length and the per-surface memory budget.
136    pub fn validate(&self) -> anyhow::Result<()> {
137        anyhow::ensure!(
138            (1..=16384).contains(&self.width) && (1..=16384).contains(&self.height),
139            "surface dimensions must be 1..=16384"
140        );
141        anyhow::ensure!(
142            (2..=8).contains(&self.slots),
143            "surface ring must hold 2..=8 slots"
144        );
145        let bytes = self.width as u64
146            * self.height as u64
147            * self.format.bytes_per_pixel() as u64
148            * self.slots as u64;
149        anyhow::ensure!(bytes <= 512 * 1024 * 1024, "surface ring exceeds 512 MiB");
150        Ok(())
151    }
152    /// Slot extent in physical pixels.
153    pub fn extent(&self) -> Size<DevicePixels> {
154        size(
155            DevicePixels(self.width as i32),
156            DevicePixels(self.height as i32),
157        )
158    }
159}
160
161/// Why `acquire` returned no slot.
162#[derive(Clone, Copy, Debug, PartialEq, Eq)]
163pub enum AcquireError {
164    /// Every slot is being written, or still referenced by retained scenes or
165    /// in-flight composition. Retry after the next presented frame.
166    WouldBlock,
167    /// The surface was invalidated (its engine device was lost). Create a new
168    /// surface on a new engine device; the last frame keeps displaying.
169    Invalidated,
170}
171impl std::fmt::Display for AcquireError {
172    fn fmt(&self, f: &mut std::fmt::Formatter<'_>) -> std::fmt::Result {
173        match self {
174            Self::WouldBlock => f.write_str("every surface slot is in use"),
175            Self::Invalidated => f.write_str("the surface was invalidated"),
176        }
177    }
178}
179impl std::error::Error for AcquireError {}
180
181/// Returned when the platform or configuration cannot support a surface.
182#[derive(Debug)]
183pub struct Unsupported(pub String);
184impl std::fmt::Display for Unsupported {
185    fn fmt(&self, f: &mut std::fmt::Formatter<'_>) -> std::fmt::Result {
186        write!(f, "engine surface unsupported: {}", self.0)
187    }
188}
189impl std::error::Error for Unsupported {}
190
191/// Counters for diagnostics and pressure tests.
192#[derive(Clone, Copy, Debug, Default, PartialEq)]
193pub struct SurfaceStats {
194    /// Slots available to `acquire`.
195    pub free: usize,
196    /// Slots acquired and not yet presented or dropped.
197    pub writing: usize,
198    /// Slots presented and still referenced (latest frame, retained scenes or
199    /// in-flight composition).
200    pub published: usize,
201    /// Total successful presents.
202    pub presents: u64,
203    /// Total `acquire` calls that returned `WouldBlock`.
204    pub would_block: u64,
205    /// Presented frames not shown yet because their sync point has not been
206    /// observed reached (see `EngineSurface::present`).
207    pub waiting: usize,
208    /// Presented frames the compositor has composited at least once.
209    pub composited: u64,
210    /// Presented frames dropped without ever being composited: the engine
211    /// produced them faster than the compositor showed them.
212    pub unshown: u64,
213    /// Delay from `present` to the first composition of the latest
214    /// composited frame. Rising values mean the compositor is falling
215    /// behind (for example, the GPU is saturated); engines can throttle.
216    pub composite_latency_ms: f32,
217}
218
219/// How engine output reaches the compositor.
220#[derive(Clone, Copy, Debug, PartialEq, Eq)]
221pub enum Transport {
222    /// The compositor samples the engine's memory directly.
223    ZeroCopy,
224}
225
226/// Engine-side helpers shared by the platform tests.
227#[cfg(all(
228    test,
229    feature = "wgpu",
230    any(target_os = "macos", target_os = "linux", target_os = "windows")
231))]
232pub(crate) mod test_engine {
233    use crate::wgpu;
234
235    /// Encode a compute pass that writes `color` to every texel of an
236    /// RGBA16F slot through a storage binding.
237    pub fn compute_fill(
238        device: &wgpu::Device,
239        texture: &wgpu::Texture,
240        color: [f32; 4],
241    ) -> wgpu::CommandBuffer {
242        let [r, g, b, a] = color;
243        let shader = device.create_shader_module(wgpu::ShaderModuleDescriptor {
244            label: Some("engine_compute_fill"),
245            source: wgpu::ShaderSource::Wgsl(
246                format!(
247                    "@group(0) @binding(0) var slot: texture_storage_2d<rgba16float, write>;
248                     @compute @workgroup_size(8, 8)
249                     fn main(@builtin(global_invocation_id) id: vec3<u32>) {{
250                         let size = textureDimensions(slot);
251                         if (id.x < size.x && id.y < size.y) {{
252                             textureStore(slot, vec2<i32>(id.xy), vec4<f32>({r:?}, {g:?}, {b:?}, {a:?}));
253                         }}
254                     }}"
255                )
256                .into(),
257            ),
258        });
259        let pipeline = device.create_compute_pipeline(&wgpu::ComputePipelineDescriptor {
260            label: Some("engine_compute_fill"),
261            layout: None,
262            module: &shader,
263            entry_point: Some("main"),
264            compilation_options: Default::default(),
265            cache: None,
266        });
267        let view = texture.create_view(&Default::default());
268        let bind = device.create_bind_group(&wgpu::BindGroupDescriptor {
269            label: Some("engine_compute_fill"),
270            layout: &pipeline.get_bind_group_layout(0),
271            entries: &[wgpu::BindGroupEntry {
272                binding: 0,
273                resource: wgpu::BindingResource::TextureView(&view),
274            }],
275        });
276        let mut encoder = device.create_command_encoder(&Default::default());
277        {
278            let mut pass = encoder.begin_compute_pass(&Default::default());
279            pass.set_pipeline(&pipeline);
280            pass.set_bind_group(0, &bind, &[]);
281            let size = texture.size();
282            pass.dispatch_workgroups(size.width.div_ceil(8), size.height.div_ceil(8), 1);
283        }
284        encoder.finish()
285    }
286}
287
288/// Independent CPU reference for compositor colour tests (#197 acceptance:
289/// near-black, saturated out-of-sRGB, fractional alpha, opacity and sRGB
290/// controls, preview and wide).
291#[cfg(test)]
292// Used by the platform tests; Linux and Windows have none without `wgpu`.
293#[cfg_attr(
294    not(any(
295        target_os = "macos",
296        all(any(target_os = "linux", target_os = "windows"), feature = "wgpu")
297    )),
298    allow(dead_code)
299)]
300pub(crate) mod colour_reference {
301    /// Stripe width in pixels; the strip is `PATCHES.len() * STRIPE` wide.
302    pub const STRIPE: u32 = 8;
303    pub const HEIGHT: u32 = 8;
304
305    /// Associated (premultiplied) linear Display P3, exactly representable
306    /// in binary16.
307    pub const PATCHES: [(&str, [f32; 4]); 5] = [
308        ("near-black", [0.001953125, 0.001953125, 0.001953125, 1.]),
309        ("P3 green (outside sRGB)", [0., 1., 0., 1.]),
310        ("P3 red (outside sRGB)", [1., 0., 0., 1.]),
311        ("half alpha", [0.25, 0.125, 0., 0.5]),
312        ("mid grey", [0.5, 0.5, 0.5, 1.]),
313    ];
314
315    /// Straight-alpha encoded sRGB BGRA8 controls (as R, G, B, A).
316    pub const SRGB_CONTROLS: [(&str, [u8; 4]); 3] = [
317        ("sRGB colour", [128, 64, 200, 255]),
318        ("sRGB near-black", [3, 3, 3, 255]),
319        ("sRGB half alpha", [255, 128, 0, 128]),
320    ];
321
322    pub fn width() -> u32 {
323        PATCHES.len() as u32 * STRIPE
324    }
325
326    /// Centre x of stripe `i`.
327    pub fn centre(i: usize) -> u32 {
328        i as u32 * STRIPE + STRIPE / 2
329    }
330
331    /// IEEE binary16 bits, round to nearest even (normal and subnormal).
332    pub fn f16_bits(value: f32) -> u16 {
333        let bits = value.to_bits();
334        let sign = ((bits >> 16) & 0x8000) as u16;
335        let exponent = ((bits >> 23) & 0xff) as i32 - 127 + 15;
336        let mantissa = bits & 0x7f_ffff;
337        if value == 0. {
338            return sign;
339        }
340        if exponent >= 31 {
341            return sign | 0x7c00;
342        }
343        if exponent <= 0 {
344            let shift = 14 - exponent;
345            let m = (mantissa | 0x80_0000) as u64;
346            let half = 1u64 << (shift - 1);
347            let rounded = (m + half - 1 + ((m >> shift) & 1)) >> shift;
348            return sign | rounded as u16;
349        }
350        let rounded = (mantissa + 0xfff + ((mantissa >> 13) & 1)) >> 13;
351        sign | (((exponent as u32) << 10) + rounded) as u16
352    }
353
354    /// The patch strip as RGBA16F rows.
355    pub fn strip_rgba16f() -> Vec<u8> {
356        let mut bytes = Vec::new();
357        for _ in 0..HEIGHT {
358            for (_, patch) in PATCHES {
359                for _ in 0..STRIPE {
360                    for channel in patch {
361                        bytes.extend_from_slice(&f16_bits(channel).to_ne_bytes());
362                    }
363                }
364            }
365        }
366        bytes
367    }
368
369    /// The sRGB controls as BGRA8 rows, one stripe each.
370    pub fn controls_bgra8() -> Vec<u8> {
371        let mut bytes = Vec::new();
372        for _ in 0..HEIGHT {
373            for (_, [r, g, b, a]) in SRGB_CONTROLS {
374                for _ in 0..STRIPE {
375                    bytes.extend_from_slice(&[b, g, r, a]);
376                }
377            }
378        }
379        bytes
380    }
381
382    /// Linear Display P3 (D65) to linear sRGB (D65), from the primaries.
383    const P3_TO_SRGB: [[f64; 3]; 3] = [
384        [1.224_940_2, -0.224_940_2, 0.],
385        [-0.042_056_95, 1.042_056_95, 0.],
386        [-0.019_637_55, -0.078_636_05, 1.098_273_6],
387    ];
388
389    fn encode(linear: f64) -> f64 {
390        let v = linear.clamp(0., 1.);
391        if v <= 0.003_130_8 {
392            v * 12.92
393        } else {
394            1.055 * v.powf(1. / 2.4) - 0.055
395        }
396    }
397
398    /// Expected 8-bit readback of a patch shown as an sRGB preview (clipped
399    /// to sRGB) at `opacity`: the encoded straight colour times coverage, as
400    /// composited over black (alpha is coverage over a transparent target).
401    pub fn preview(patch: [f32; 4], opacity: f32) -> [f64; 4] {
402        let alpha = patch[3] as f64;
403        let straight: Vec<f64> = (0..3)
404            .map(|c| {
405                if alpha > 0. {
406                    patch[c] as f64 / alpha
407                } else {
408                    0.
409                }
410            })
411            .collect();
412        let coverage = alpha * opacity as f64;
413        let mut out = [0.; 4];
414        for (row, value) in P3_TO_SRGB.iter().zip(out.iter_mut()) {
415            let linear: f64 = row.iter().zip(&straight).map(|(m, v)| m * v).sum();
416            *value = encode(linear) * 255. * coverage;
417        }
418        out[3] = coverage * 255.;
419        out
420    }
421
422    /// Expected 8-bit readback of a straight sRGB control at `opacity`.
423    pub fn control(rgba: [u8; 4], opacity: f32) -> [f64; 4] {
424        let coverage = rgba[3] as f64 / 255. * opacity as f64;
425        [
426            rgba[0] as f64 * coverage,
427            rgba[1] as f64 * coverage,
428            rgba[2] as f64 * coverage,
429            coverage * 255.,
430        ]
431    }
432
433    /// Max colour-channel distance between a readback pixel and a reference.
434    /// Headless compositors draw onto the window background (opaque black on
435    /// macOS), so alpha is not compared; associated colour over black equals
436    /// the reference either way.
437    pub fn distance(actual: [u8; 4], expected: [f64; 4]) -> f64 {
438        (0..3)
439            .map(|c| (actual[c] as f64 - expected[c]).abs())
440            .fold(0., f64::max)
441    }
442}
443
444/// binary16 bits to f32 (macOS FP16 composition readbacks).
445#[cfg(all(test, target_os = "macos"))]
446pub(crate) fn colour_reference_f16(bits: u16) -> f32 {
447    let sign = if bits & 0x8000 != 0 { -1. } else { 1. };
448    let exponent = ((bits >> 10) & 0x1f) as i32;
449    let mantissa = (bits & 0x3ff) as f32;
450    sign * match exponent {
451        0 => mantissa * 2f32.powi(-24),
452        31 => f32::INFINITY,
453        _ => (1. + mantissa / 1024.) * 2f32.powi(exponent - 15),
454    }
455}
456
457#[cfg(test)]
458mod tests {
459    use super::*;
460    #[test]
461    fn validates_descriptions() {
462        let ok = SurfaceDesc {
463            width: 64,
464            height: 32,
465            format: SurfaceFormat::Rgba16FloatLinearP3,
466            slots: 3,
467        };
468        assert!(ok.validate().is_ok());
469        for bad in [
470            SurfaceDesc { width: 0, ..ok },
471            SurfaceDesc {
472                height: 16385,
473                ..ok
474            },
475            SurfaceDesc { slots: 1, ..ok },
476            SurfaceDesc { slots: 9, ..ok },
477            SurfaceDesc {
478                width: 16384,
479                height: 16384,
480                ..ok
481            },
482        ] {
483            assert!(bad.validate().is_err(), "{bad:?}");
484        }
485    }
486}
xgpui_gpu_interop/lib.rs

xgpui_gpu_interop/
lib.rs
