From ac83f79d94cd8ab2c8186e5344a9225831fd7505 Mon Sep 17 00:00:00 2001 From: Patrick Walton Date: Thu, 23 Apr 2020 18:00:52 -0700 Subject: [PATCH] Add a compute shader path, optimize GPU memory management, and switch from SDL to `surfman`. This is a large commit; explanations of each change follow. This adds an optional compute shader path, off by default, for rendering fills to alpha masks. It usually does not improve performance at present, but it provides a good baseline for further optimizations. Later improvements will likely aim to avoid writes to the mask texture entirely. Supporting infrastructure for compute shader has been added to `pathfinder_gpu` for the OpenGL and Metal backends. The Metal backend has been optimized to avoid unneccessary buffer allocations and reflection. As part of this, argument buffers have been removed, as the current SPIRV-Cross compiler no longer requires them. The GPU renderer has been improved to avoid stalls. Now, separate buffers are allocated for each fill batch and for each frame. This can be extended in the future to allow for separate buffers for tile draw operations as well. SDL usage has been removed in favor of the native Rust `surfman` and `winit`. Because `surfman` allows for selection of the integrated GPU on multi-GPU system, it is chosen by default. The demo supports a new `--high-performance-gpu` option to opt into the discrete GPU. --- Cargo.lock | 376 ++++++- c/src/lib.rs | 14 +- demo/common/Cargo.toml | 3 + demo/common/src/device.rs | 2 +- demo/common/src/lib.rs | 40 +- demo/common/src/renderer.rs | 6 +- demo/common/src/window.rs | 12 +- demo/native/Cargo.toml | 17 +- demo/native/src/main.rs | 537 ++++++---- examples/canvas_glutin_minimal/src/main.rs | 5 +- examples/canvas_minimal/Cargo.toml | 13 +- examples/canvas_minimal/src/main.rs | 102 +- examples/canvas_moire/Cargo.toml | 14 +- examples/canvas_moire/src/main.rs | 110 ++- examples/canvas_nanovg/Cargo.toml | 14 +- examples/canvas_nanovg/src/main.rs | 111 ++- examples/canvas_text/src/main.rs | 5 +- examples/swf_basic/src/main.rs | 5 +- gl/src/lib.rs | 177 +++- gpu/src/lib.rs | 111 ++- metal/Cargo.toml | 4 +- metal/src/lib.rs | 924 +++++++++++++----- renderer/src/allocator.rs | 1 + renderer/src/builder.rs | 25 - renderer/src/gpu/options.rs | 6 +- renderer/src/gpu/renderer.rs | 796 ++++++++++----- renderer/src/gpu/shaders.rs | 147 +-- renderer/src/gpu_data.rs | 5 +- resources/shaders/gl3/fill.fs.glsl | 29 +- resources/shaders/metal/blit.fs.metal | 10 +- resources/shaders/metal/debug_solid.fs.metal | 9 +- resources/shaders/metal/debug_solid.vs.metal | 9 +- .../shaders/metal/debug_texture.fs.metal | 13 +- .../shaders/metal/debug_texture.vs.metal | 12 +- resources/shaders/metal/demo_ground.fs.metal | 10 +- resources/shaders/metal/demo_ground.vs.metal | 12 +- resources/shaders/metal/fill.cs.metal | 67 ++ resources/shaders/metal/fill.fs.metal | 24 +- resources/shaders/metal/fill.vs.metal | 17 +- resources/shaders/metal/reproject.fs.metal | 13 +- resources/shaders/metal/reproject.vs.metal | 9 +- resources/shaders/metal/tile.fs.metal | 51 +- resources/shaders/metal/tile.vs.metal | 23 +- resources/shaders/metal/tile_clip.fs.metal | 10 +- resources/shaders/metal/tile_copy.fs.metal | 13 +- resources/shaders/metal/tile_copy.vs.metal | 12 +- shaders/Makefile | 13 +- shaders/fill.cs.glsl | 67 ++ shaders/fill.fs.glsl | 27 +- shaders/fill.inc.glsl | 27 + shaders/fill.vs.glsl | 2 +- ui/src/lib.rs | 24 +- webgl/src/lib.rs | 74 +- 53 files changed, 2975 insertions(+), 1184 deletions(-) create mode 100644 resources/shaders/metal/fill.cs.metal create mode 100644 shaders/fill.cs.glsl create mode 100644 shaders/fill.inc.glsl diff --git a/Cargo.lock b/Cargo.lock index 34382207..0f7f62a8 100644 --- a/Cargo.lock +++ b/Cargo.lock @@ -233,6 +233,7 @@ dependencies = [ name = "canvas_minimal" version = "0.1.0" dependencies = [ + "euclid 0.20.10 (registry+https://github.com/rust-lang/crates.io-index)", "gl 0.14.0 (registry+https://github.com/rust-lang/crates.io-index)", "pathfinder_canvas 0.5.0", "pathfinder_color 0.5.0", @@ -242,14 +243,15 @@ dependencies = [ "pathfinder_gpu 0.5.0", "pathfinder_renderer 0.5.0", "pathfinder_resources 0.5.0", - "sdl2 0.33.0 (registry+https://github.com/rust-lang/crates.io-index)", - "sdl2-sys 0.33.0 (registry+https://github.com/rust-lang/crates.io-index)", + "surfman 0.2.0 (registry+https://github.com/rust-lang/crates.io-index)", + "winit 0.19.3 (registry+https://github.com/rust-lang/crates.io-index)", ] [[package]] name = "canvas_moire" version = "0.1.0" dependencies = [ + "euclid 0.20.10 (registry+https://github.com/rust-lang/crates.io-index)", "gl 0.14.0 (registry+https://github.com/rust-lang/crates.io-index)", "pathfinder_canvas 0.5.0", "pathfinder_color 0.5.0", @@ -259,8 +261,8 @@ dependencies = [ "pathfinder_gpu 0.5.0", "pathfinder_renderer 0.5.0", "pathfinder_resources 0.5.0", - "sdl2 0.33.0 (registry+https://github.com/rust-lang/crates.io-index)", - "sdl2-sys 0.33.0 (registry+https://github.com/rust-lang/crates.io-index)", + "surfman 0.2.0", + "winit 0.19.3 (registry+https://github.com/rust-lang/crates.io-index)", ] [[package]] @@ -268,6 +270,7 @@ name = "canvas_nanovg" version = "0.1.0" dependencies = [ "arrayvec 0.5.1 (registry+https://github.com/rust-lang/crates.io-index)", + "euclid 0.20.10 (registry+https://github.com/rust-lang/crates.io-index)", "font-kit 0.6.0 (registry+https://github.com/rust-lang/crates.io-index)", "gl 0.14.0 (registry+https://github.com/rust-lang/crates.io-index)", "image 0.23.3 (registry+https://github.com/rust-lang/crates.io-index)", @@ -282,8 +285,8 @@ dependencies = [ "pathfinder_renderer 0.5.0", "pathfinder_resources 0.5.0", "pathfinder_simd 0.5.0", - "sdl2 0.33.0 (registry+https://github.com/rust-lang/crates.io-index)", - "sdl2-sys 0.33.0 (registry+https://github.com/rust-lang/crates.io-index)", + "surfman 0.2.0 (git+https://github.com/pcwalton/surfman?rev=bc29e7ae88ca7dd64d3b1c7e185604693290207f)", + "winit 0.19.3 (registry+https://github.com/rust-lang/crates.io-index)", ] [[package]] @@ -373,6 +376,20 @@ dependencies = [ "cc 1.0.50 (registry+https://github.com/rust-lang/crates.io-index)", ] +[[package]] +name = "cocoa" +version = "0.18.5" +source = "registry+https://github.com/rust-lang/crates.io-index" +dependencies = [ + "bitflags 1.2.1 (registry+https://github.com/rust-lang/crates.io-index)", + "block 0.1.6 (registry+https://github.com/rust-lang/crates.io-index)", + "core-foundation 0.6.4 (registry+https://github.com/rust-lang/crates.io-index)", + "core-graphics 0.17.3 (registry+https://github.com/rust-lang/crates.io-index)", + "foreign-types 0.3.2 (registry+https://github.com/rust-lang/crates.io-index)", + "libc 0.2.69 (registry+https://github.com/rust-lang/crates.io-index)", + "objc 0.2.7 (registry+https://github.com/rust-lang/crates.io-index)", +] + [[package]] name = "cocoa" version = "0.19.1" @@ -620,11 +637,15 @@ name = "demo" version = "0.1.0" dependencies = [ "color-backtrace 0.3.0 (registry+https://github.com/rust-lang/crates.io-index)", + "euclid 0.20.10 (registry+https://github.com/rust-lang/crates.io-index)", "foreign-types 0.3.2 (registry+https://github.com/rust-lang/crates.io-index)", "gl 0.14.0 (registry+https://github.com/rust-lang/crates.io-index)", + "io-surface 0.12.1 (registry+https://github.com/rust-lang/crates.io-index)", "jemallocator 0.3.2 (registry+https://github.com/rust-lang/crates.io-index)", + "lazy_static 1.4.0 (registry+https://github.com/rust-lang/crates.io-index)", "metal 0.17.1 (registry+https://github.com/rust-lang/crates.io-index)", "nfd 0.0.4 (registry+https://github.com/rust-lang/crates.io-index)", + "objc 0.2.7 (registry+https://github.com/rust-lang/crates.io-index)", "pathfinder_demo 0.1.0", "pathfinder_geometry 0.5.1", "pathfinder_gl 0.5.0", @@ -633,8 +654,8 @@ dependencies = [ "pathfinder_resources 0.5.0", "pathfinder_simd 0.5.0", "pretty_env_logger 0.4.0 (registry+https://github.com/rust-lang/crates.io-index)", - "sdl2 0.33.0 (registry+https://github.com/rust-lang/crates.io-index)", - "sdl2-sys 0.33.0 (registry+https://github.com/rust-lang/crates.io-index)", + "surfman 0.2.0", + "winit 0.19.3 (registry+https://github.com/rust-lang/crates.io-index)", ] [[package]] @@ -662,6 +683,17 @@ name = "dispatch" version = "0.2.0" source = "registry+https://github.com/rust-lang/crates.io-index" +[[package]] +name = "display-link" +version = "0.2.0" +source = "registry+https://github.com/rust-lang/crates.io-index" +dependencies = [ + "foreign-types 0.3.2 (registry+https://github.com/rust-lang/crates.io-index)", + "objc 0.2.7 (registry+https://github.com/rust-lang/crates.io-index)", + "objc-foundation 0.1.1 (registry+https://github.com/rust-lang/crates.io-index)", + "time-point 0.1.1 (registry+https://github.com/rust-lang/crates.io-index)", +] + [[package]] name = "dlib" version = "0.4.1" @@ -899,6 +931,16 @@ dependencies = [ "gl_generator 0.14.0 (registry+https://github.com/rust-lang/crates.io-index)", ] +[[package]] +name = "gl_generator" +version = "0.11.0" +source = "registry+https://github.com/rust-lang/crates.io-index" +dependencies = [ + "khronos_api 3.1.0 (registry+https://github.com/rust-lang/crates.io-index)", + "log 0.4.8 (registry+https://github.com/rust-lang/crates.io-index)", + "xml-rs 0.8.2 (registry+https://github.com/rust-lang/crates.io-index)", +] + [[package]] name = "gl_generator" version = "0.13.1" @@ -919,6 +961,14 @@ dependencies = [ "xml-rs 0.8.2 (registry+https://github.com/rust-lang/crates.io-index)", ] +[[package]] +name = "gleam" +version = "0.7.0" +source = "registry+https://github.com/rust-lang/crates.io-index" +dependencies = [ + "gl_generator 0.13.1 (registry+https://github.com/rust-lang/crates.io-index)", +] + [[package]] name = "glutin" version = "0.23.0" @@ -1094,6 +1144,18 @@ dependencies = [ "web-sys 0.3.37 (registry+https://github.com/rust-lang/crates.io-index)", ] +[[package]] +name = "io-surface" +version = "0.12.1" +source = "registry+https://github.com/rust-lang/crates.io-index" +dependencies = [ + "cgl 0.3.2 (registry+https://github.com/rust-lang/crates.io-index)", + "core-foundation 0.6.4 (registry+https://github.com/rust-lang/crates.io-index)", + "gleam 0.7.0 (registry+https://github.com/rust-lang/crates.io-index)", + "leaky-cow 0.1.1 (registry+https://github.com/rust-lang/crates.io-index)", + "libc 0.2.69 (registry+https://github.com/rust-lang/crates.io-index)", +] + [[package]] name = "iovec" version = "0.1.4" @@ -1201,6 +1263,19 @@ name = "lazycell" version = "1.2.1" source = "registry+https://github.com/rust-lang/crates.io-index" +[[package]] +name = "leak" +version = "0.1.2" +source = "registry+https://github.com/rust-lang/crates.io-index" + +[[package]] +name = "leaky-cow" +version = "0.1.1" +source = "registry+https://github.com/rust-lang/crates.io-index" +dependencies = [ + "leak 0.1.2 (registry+https://github.com/rust-lang/crates.io-index)", +] + [[package]] name = "lexical-core" version = "0.6.2" @@ -1278,6 +1353,14 @@ name = "lzw" version = "0.10.0" source = "registry+https://github.com/rust-lang/crates.io-index" +[[package]] +name = "mach" +version = "0.3.2" +source = "registry+https://github.com/rust-lang/crates.io-index" +dependencies = [ + "libc 0.2.69 (registry+https://github.com/rust-lang/crates.io-index)", +] + [[package]] name = "malloc_buf" version = "0.0.6" @@ -1488,6 +1571,16 @@ dependencies = [ "objc_exception 0.1.2 (registry+https://github.com/rust-lang/crates.io-index)", ] +[[package]] +name = "objc-foundation" +version = "0.1.1" +source = "registry+https://github.com/rust-lang/crates.io-index" +dependencies = [ + "block 0.1.6 (registry+https://github.com/rust-lang/crates.io-index)", + "objc 0.2.7 (registry+https://github.com/rust-lang/crates.io-index)", + "objc_id 0.1.1 (registry+https://github.com/rust-lang/crates.io-index)", +] + [[package]] name = "objc_exception" version = "0.1.2" @@ -1496,6 +1589,14 @@ dependencies = [ "cc 1.0.50 (registry+https://github.com/rust-lang/crates.io-index)", ] +[[package]] +name = "objc_id" +version = "0.1.1" +source = "registry+https://github.com/rust-lang/crates.io-index" +dependencies = [ + "objc 0.2.7 (registry+https://github.com/rust-lang/crates.io-index)", +] + [[package]] name = "ordered-float" version = "1.0.2" @@ -1512,6 +1613,16 @@ dependencies = [ "shared_library 0.1.9 (registry+https://github.com/rust-lang/crates.io-index)", ] +[[package]] +name = "parking_lot" +version = "0.9.0" +source = "registry+https://github.com/rust-lang/crates.io-index" +dependencies = [ + "lock_api 0.3.4 (registry+https://github.com/rust-lang/crates.io-index)", + "parking_lot_core 0.6.2 (registry+https://github.com/rust-lang/crates.io-index)", + "rustc_version 0.2.3 (registry+https://github.com/rust-lang/crates.io-index)", +] + [[package]] name = "parking_lot" version = "0.10.2" @@ -1521,6 +1632,20 @@ dependencies = [ "parking_lot_core 0.7.1 (registry+https://github.com/rust-lang/crates.io-index)", ] +[[package]] +name = "parking_lot_core" +version = "0.6.2" +source = "registry+https://github.com/rust-lang/crates.io-index" +dependencies = [ + "cfg-if 0.1.10 (registry+https://github.com/rust-lang/crates.io-index)", + "cloudabi 0.0.3 (registry+https://github.com/rust-lang/crates.io-index)", + "libc 0.2.69 (registry+https://github.com/rust-lang/crates.io-index)", + "redox_syscall 0.1.56 (registry+https://github.com/rust-lang/crates.io-index)", + "rustc_version 0.2.3 (registry+https://github.com/rust-lang/crates.io-index)", + "smallvec 0.6.13 (registry+https://github.com/rust-lang/crates.io-index)", + "winapi 0.3.8 (registry+https://github.com/rust-lang/crates.io-index)", +] + [[package]] name = "parking_lot_core" version = "0.7.1" @@ -1612,6 +1737,7 @@ dependencies = [ "clap 2.33.0 (registry+https://github.com/rust-lang/crates.io-index)", "gl 0.14.0 (registry+https://github.com/rust-lang/crates.io-index)", "image 0.23.3 (registry+https://github.com/rust-lang/crates.io-index)", + "io-surface 0.12.1 (registry+https://github.com/rust-lang/crates.io-index)", "log 0.4.8 (registry+https://github.com/rust-lang/crates.io-index)", "metal 0.17.1 (registry+https://github.com/rust-lang/crates.io-index)", "pathfinder_color 0.5.0", @@ -1715,9 +1841,11 @@ dependencies = [ "block 0.1.6 (registry+https://github.com/rust-lang/crates.io-index)", "byteorder 1.3.4 (registry+https://github.com/rust-lang/crates.io-index)", "cocoa 0.19.1 (registry+https://github.com/rust-lang/crates.io-index)", - "core-foundation 0.7.0 (registry+https://github.com/rust-lang/crates.io-index)", + "core-foundation 0.6.4 (registry+https://github.com/rust-lang/crates.io-index)", "foreign-types 0.3.2 (registry+https://github.com/rust-lang/crates.io-index)", "half 1.5.0 (registry+https://github.com/rust-lang/crates.io-index)", + "io-surface 0.12.1 (registry+https://github.com/rust-lang/crates.io-index)", + "libc 0.2.69 (registry+https://github.com/rust-lang/crates.io-index)", "metal 0.17.1 (registry+https://github.com/rust-lang/crates.io-index)", "objc 0.2.7 (registry+https://github.com/rust-lang/crates.io-index)", "pathfinder_geometry 0.5.1", @@ -2258,11 +2386,35 @@ name = "slab" version = "0.4.2" source = "registry+https://github.com/rust-lang/crates.io-index" +[[package]] +name = "smallvec" +version = "0.6.13" +source = "registry+https://github.com/rust-lang/crates.io-index" +dependencies = [ + "maybe-uninit 2.0.0 (registry+https://github.com/rust-lang/crates.io-index)", +] + [[package]] name = "smallvec" version = "1.3.0" source = "registry+https://github.com/rust-lang/crates.io-index" +[[package]] +name = "smithay-client-toolkit" +version = "0.4.6" +source = "registry+https://github.com/rust-lang/crates.io-index" +dependencies = [ + "andrew 0.2.1 (registry+https://github.com/rust-lang/crates.io-index)", + "bitflags 1.2.1 (registry+https://github.com/rust-lang/crates.io-index)", + "dlib 0.4.1 (registry+https://github.com/rust-lang/crates.io-index)", + "lazy_static 1.4.0 (registry+https://github.com/rust-lang/crates.io-index)", + "memmap 0.7.0 (registry+https://github.com/rust-lang/crates.io-index)", + "nix 0.14.1 (registry+https://github.com/rust-lang/crates.io-index)", + "wayland-client 0.21.13 (registry+https://github.com/rust-lang/crates.io-index)", + "wayland-commons 0.21.13 (registry+https://github.com/rust-lang/crates.io-index)", + "wayland-protocols 0.21.13 (registry+https://github.com/rust-lang/crates.io-index)", +] + [[package]] name = "smithay-client-toolkit" version = "0.6.6" @@ -2301,6 +2453,87 @@ name = "strsim" version = "0.8.0" source = "registry+https://github.com/rust-lang/crates.io-index" +[[package]] +name = "surfman" +version = "0.2.0" +source = "git+https://github.com/pcwalton/surfman?rev=bc29e7ae88ca7dd64d3b1c7e185604693290207f#bc29e7ae88ca7dd64d3b1c7e185604693290207f" +dependencies = [ + "bitflags 1.2.1 (registry+https://github.com/rust-lang/crates.io-index)", + "cgl 0.3.2 (registry+https://github.com/rust-lang/crates.io-index)", + "cocoa 0.19.1 (registry+https://github.com/rust-lang/crates.io-index)", + "core-foundation 0.6.4 (registry+https://github.com/rust-lang/crates.io-index)", + "core-graphics 0.17.3 (registry+https://github.com/rust-lang/crates.io-index)", + "display-link 0.2.0 (registry+https://github.com/rust-lang/crates.io-index)", + "euclid 0.20.10 (registry+https://github.com/rust-lang/crates.io-index)", + "gl_generator 0.11.0 (registry+https://github.com/rust-lang/crates.io-index)", + "io-surface 0.12.1 (registry+https://github.com/rust-lang/crates.io-index)", + "lazy_static 1.4.0 (registry+https://github.com/rust-lang/crates.io-index)", + "libc 0.2.69 (registry+https://github.com/rust-lang/crates.io-index)", + "log 0.4.8 (registry+https://github.com/rust-lang/crates.io-index)", + "mach 0.3.2 (registry+https://github.com/rust-lang/crates.io-index)", + "objc 0.2.7 (registry+https://github.com/rust-lang/crates.io-index)", + "parking_lot 0.9.0 (registry+https://github.com/rust-lang/crates.io-index)", + "wayland-sys 0.24.1 (registry+https://github.com/rust-lang/crates.io-index)", + "winapi 0.3.8 (registry+https://github.com/rust-lang/crates.io-index)", + "winit 0.19.3 (registry+https://github.com/rust-lang/crates.io-index)", + "wio 0.2.2 (registry+https://github.com/rust-lang/crates.io-index)", + "x11 2.18.2 (registry+https://github.com/rust-lang/crates.io-index)", +] + +[[package]] +name = "surfman" +version = "0.2.0" +dependencies = [ + "bitflags 1.2.1 (registry+https://github.com/rust-lang/crates.io-index)", + "cgl 0.3.2 (registry+https://github.com/rust-lang/crates.io-index)", + "cocoa 0.19.1 (registry+https://github.com/rust-lang/crates.io-index)", + "core-foundation 0.6.4 (registry+https://github.com/rust-lang/crates.io-index)", + "core-graphics 0.17.3 (registry+https://github.com/rust-lang/crates.io-index)", + "display-link 0.2.0 (registry+https://github.com/rust-lang/crates.io-index)", + "euclid 0.20.10 (registry+https://github.com/rust-lang/crates.io-index)", + "gl_generator 0.11.0 (registry+https://github.com/rust-lang/crates.io-index)", + "io-surface 0.12.1 (registry+https://github.com/rust-lang/crates.io-index)", + "lazy_static 1.4.0 (registry+https://github.com/rust-lang/crates.io-index)", + "libc 0.2.69 (registry+https://github.com/rust-lang/crates.io-index)", + "log 0.4.8 (registry+https://github.com/rust-lang/crates.io-index)", + "mach 0.3.2 (registry+https://github.com/rust-lang/crates.io-index)", + "metal 0.17.1 (registry+https://github.com/rust-lang/crates.io-index)", + "objc 0.2.7 (registry+https://github.com/rust-lang/crates.io-index)", + "parking_lot 0.9.0 (registry+https://github.com/rust-lang/crates.io-index)", + "wayland-sys 0.24.1 (registry+https://github.com/rust-lang/crates.io-index)", + "winapi 0.3.8 (registry+https://github.com/rust-lang/crates.io-index)", + "winit 0.19.3 (registry+https://github.com/rust-lang/crates.io-index)", + "wio 0.2.2 (registry+https://github.com/rust-lang/crates.io-index)", + "x11 2.18.2 (registry+https://github.com/rust-lang/crates.io-index)", +] + +[[package]] +name = "surfman" +version = "0.2.0" +source = "registry+https://github.com/rust-lang/crates.io-index" +dependencies = [ + "bitflags 1.2.1 (registry+https://github.com/rust-lang/crates.io-index)", + "cgl 0.3.2 (registry+https://github.com/rust-lang/crates.io-index)", + "cocoa 0.19.1 (registry+https://github.com/rust-lang/crates.io-index)", + "core-foundation 0.6.4 (registry+https://github.com/rust-lang/crates.io-index)", + "core-graphics 0.17.3 (registry+https://github.com/rust-lang/crates.io-index)", + "display-link 0.2.0 (registry+https://github.com/rust-lang/crates.io-index)", + "euclid 0.20.10 (registry+https://github.com/rust-lang/crates.io-index)", + "gl_generator 0.11.0 (registry+https://github.com/rust-lang/crates.io-index)", + "io-surface 0.12.1 (registry+https://github.com/rust-lang/crates.io-index)", + "lazy_static 1.4.0 (registry+https://github.com/rust-lang/crates.io-index)", + "libc 0.2.69 (registry+https://github.com/rust-lang/crates.io-index)", + "log 0.4.8 (registry+https://github.com/rust-lang/crates.io-index)", + "mach 0.3.2 (registry+https://github.com/rust-lang/crates.io-index)", + "objc 0.2.7 (registry+https://github.com/rust-lang/crates.io-index)", + "parking_lot 0.9.0 (registry+https://github.com/rust-lang/crates.io-index)", + "wayland-sys 0.24.1 (registry+https://github.com/rust-lang/crates.io-index)", + "winapi 0.3.8 (registry+https://github.com/rust-lang/crates.io-index)", + "winit 0.19.3 (registry+https://github.com/rust-lang/crates.io-index)", + "wio 0.2.2 (registry+https://github.com/rust-lang/crates.io-index)", + "x11 2.18.2 (registry+https://github.com/rust-lang/crates.io-index)", +] + [[package]] name = "svg-to-skia" version = "0.1.0" @@ -2423,6 +2656,11 @@ dependencies = [ "miniz_oxide 0.3.6 (registry+https://github.com/rust-lang/crates.io-index)", ] +[[package]] +name = "time-point" +version = "0.1.1" +source = "registry+https://github.com/rust-lang/crates.io-index" + [[package]] name = "toml" version = "0.5.6" @@ -2588,6 +2826,20 @@ name = "wasm-bindgen-shared" version = "0.2.60" source = "registry+https://github.com/rust-lang/crates.io-index" +[[package]] +name = "wayland-client" +version = "0.21.13" +source = "registry+https://github.com/rust-lang/crates.io-index" +dependencies = [ + "bitflags 1.2.1 (registry+https://github.com/rust-lang/crates.io-index)", + "downcast-rs 1.1.1 (registry+https://github.com/rust-lang/crates.io-index)", + "libc 0.2.69 (registry+https://github.com/rust-lang/crates.io-index)", + "nix 0.14.1 (registry+https://github.com/rust-lang/crates.io-index)", + "wayland-commons 0.21.13 (registry+https://github.com/rust-lang/crates.io-index)", + "wayland-scanner 0.21.13 (registry+https://github.com/rust-lang/crates.io-index)", + "wayland-sys 0.21.13 (registry+https://github.com/rust-lang/crates.io-index)", +] + [[package]] name = "wayland-client" version = "0.23.6" @@ -2604,6 +2856,15 @@ dependencies = [ "wayland-sys 0.23.6 (registry+https://github.com/rust-lang/crates.io-index)", ] +[[package]] +name = "wayland-commons" +version = "0.21.13" +source = "registry+https://github.com/rust-lang/crates.io-index" +dependencies = [ + "nix 0.14.1 (registry+https://github.com/rust-lang/crates.io-index)", + "wayland-sys 0.21.13 (registry+https://github.com/rust-lang/crates.io-index)", +] + [[package]] name = "wayland-commons" version = "0.23.6" @@ -2613,6 +2874,18 @@ dependencies = [ "wayland-sys 0.23.6 (registry+https://github.com/rust-lang/crates.io-index)", ] +[[package]] +name = "wayland-protocols" +version = "0.21.13" +source = "registry+https://github.com/rust-lang/crates.io-index" +dependencies = [ + "bitflags 1.2.1 (registry+https://github.com/rust-lang/crates.io-index)", + "wayland-client 0.21.13 (registry+https://github.com/rust-lang/crates.io-index)", + "wayland-commons 0.21.13 (registry+https://github.com/rust-lang/crates.io-index)", + "wayland-scanner 0.21.13 (registry+https://github.com/rust-lang/crates.io-index)", + "wayland-sys 0.21.13 (registry+https://github.com/rust-lang/crates.io-index)", +] + [[package]] name = "wayland-protocols" version = "0.23.6" @@ -2624,6 +2897,16 @@ dependencies = [ "wayland-scanner 0.23.6 (registry+https://github.com/rust-lang/crates.io-index)", ] +[[package]] +name = "wayland-scanner" +version = "0.21.13" +source = "registry+https://github.com/rust-lang/crates.io-index" +dependencies = [ + "proc-macro2 0.4.30 (registry+https://github.com/rust-lang/crates.io-index)", + "quote 0.6.13 (registry+https://github.com/rust-lang/crates.io-index)", + "xml-rs 0.8.2 (registry+https://github.com/rust-lang/crates.io-index)", +] + [[package]] name = "wayland-scanner" version = "0.23.6" @@ -2634,6 +2917,15 @@ dependencies = [ "xml-rs 0.8.2 (registry+https://github.com/rust-lang/crates.io-index)", ] +[[package]] +name = "wayland-sys" +version = "0.21.13" +source = "registry+https://github.com/rust-lang/crates.io-index" +dependencies = [ + "dlib 0.4.1 (registry+https://github.com/rust-lang/crates.io-index)", + "lazy_static 1.4.0 (registry+https://github.com/rust-lang/crates.io-index)", +] + [[package]] name = "wayland-sys" version = "0.23.6" @@ -2643,6 +2935,15 @@ dependencies = [ "lazy_static 1.4.0 (registry+https://github.com/rust-lang/crates.io-index)", ] +[[package]] +name = "wayland-sys" +version = "0.24.1" +source = "registry+https://github.com/rust-lang/crates.io-index" +dependencies = [ + "dlib 0.4.1 (registry+https://github.com/rust-lang/crates.io-index)", + "lazy_static 1.4.0 (registry+https://github.com/rust-lang/crates.io-index)", +] + [[package]] name = "web-sys" version = "0.3.37" @@ -2689,6 +2990,29 @@ name = "winapi-x86_64-pc-windows-gnu" version = "0.4.0" source = "registry+https://github.com/rust-lang/crates.io-index" +[[package]] +name = "winit" +version = "0.19.3" +source = "registry+https://github.com/rust-lang/crates.io-index" +dependencies = [ + "android_glue 0.2.3 (registry+https://github.com/rust-lang/crates.io-index)", + "backtrace 0.3.46 (registry+https://github.com/rust-lang/crates.io-index)", + "bitflags 1.2.1 (registry+https://github.com/rust-lang/crates.io-index)", + "cocoa 0.18.5 (registry+https://github.com/rust-lang/crates.io-index)", + "core-foundation 0.6.4 (registry+https://github.com/rust-lang/crates.io-index)", + "core-graphics 0.17.3 (registry+https://github.com/rust-lang/crates.io-index)", + "lazy_static 1.4.0 (registry+https://github.com/rust-lang/crates.io-index)", + "libc 0.2.69 (registry+https://github.com/rust-lang/crates.io-index)", + "log 0.4.8 (registry+https://github.com/rust-lang/crates.io-index)", + "objc 0.2.7 (registry+https://github.com/rust-lang/crates.io-index)", + "parking_lot 0.9.0 (registry+https://github.com/rust-lang/crates.io-index)", + "percent-encoding 2.1.0 (registry+https://github.com/rust-lang/crates.io-index)", + "smithay-client-toolkit 0.4.6 (registry+https://github.com/rust-lang/crates.io-index)", + "wayland-client 0.21.13 (registry+https://github.com/rust-lang/crates.io-index)", + "winapi 0.3.8 (registry+https://github.com/rust-lang/crates.io-index)", + "x11-dl 2.18.5 (registry+https://github.com/rust-lang/crates.io-index)", +] + [[package]] name = "winit" version = "0.21.0" @@ -2734,6 +3058,15 @@ dependencies = [ "winapi-build 0.1.1 (registry+https://github.com/rust-lang/crates.io-index)", ] +[[package]] +name = "x11" +version = "2.18.2" +source = "registry+https://github.com/rust-lang/crates.io-index" +dependencies = [ + "libc 0.2.69 (registry+https://github.com/rust-lang/crates.io-index)", + "pkg-config 0.3.17 (registry+https://github.com/rust-lang/crates.io-index)", +] + [[package]] name = "x11-dl" version = "2.18.5" @@ -2799,6 +3132,7 @@ source = "registry+https://github.com/rust-lang/crates.io-index" "checksum clap 2.33.0 (registry+https://github.com/rust-lang/crates.io-index)" = "5067f5bb2d80ef5d68b4c87db81601f0b75bca627bc2ef76b141d7b846a3c6d9" "checksum cloudabi 0.0.3 (registry+https://github.com/rust-lang/crates.io-index)" = "ddfc5b9aa5d4507acaf872de71051dfd0e309860e88966e1051e462a077aac4f" "checksum cmake 0.1.42 (registry+https://github.com/rust-lang/crates.io-index)" = "81fb25b677f8bf1eb325017cb6bb8452f87969db0fedb4f757b297bee78a7c62" +"checksum cocoa 0.18.5 (registry+https://github.com/rust-lang/crates.io-index)" = "1706996401131526e36b3b49f0c4d912639ce110996f3ca144d78946727bce54" "checksum cocoa 0.19.1 (registry+https://github.com/rust-lang/crates.io-index)" = "f29f7768b2d1be17b96158e3285951d366b40211320fb30826a76cb7a0da6400" "checksum color-backtrace 0.3.0 (registry+https://github.com/rust-lang/crates.io-index)" = "65d13f1078cc63c791d0deba0dd43db37c9ec02b311f10bed10b577016f3a957" "checksum color_quant 1.0.1 (registry+https://github.com/rust-lang/crates.io-index)" = "0dbbb57365263e881e805dc77d94697c9118fd94d8da011240555aa7b23445bd" @@ -2827,6 +3161,7 @@ source = "registry+https://github.com/rust-lang/crates.io-index" "checksum dirs 2.0.2 (registry+https://github.com/rust-lang/crates.io-index)" = "13aea89a5c93364a98e9b37b2fa237effbb694d5cfe01c5b70941f7eb087d5e3" "checksum dirs-sys 0.3.4 (registry+https://github.com/rust-lang/crates.io-index)" = "afa0b23de8fd801745c471deffa6e12d248f962c9fd4b4c33787b055599bde7b" "checksum dispatch 0.2.0 (registry+https://github.com/rust-lang/crates.io-index)" = "bd0c93bb4b0c6d9b77f4435b0ae98c24d17f1c45b2ff844c6151a07256ca923b" +"checksum display-link 0.2.0 (registry+https://github.com/rust-lang/crates.io-index)" = "303de632386f9c82eb7823456f5932bd40b4de9521078901767bf16a9f331491" "checksum dlib 0.4.1 (registry+https://github.com/rust-lang/crates.io-index)" = "77e51249a9d823a4cb79e3eca6dcd756153e8ed0157b6c04775d04bf1b13b76a" "checksum downcast-rs 1.1.1 (registry+https://github.com/rust-lang/crates.io-index)" = "52ba6eb47c2131e784a38b726eb54c1e1484904f013e576a25354d0124161af6" "checksum dwrote 0.10.0 (registry+https://github.com/rust-lang/crates.io-index)" = "bcdf488e3a52a7aa30a05732a3e58420e22acb4b2b75635a561fc6ffbcab59ef" @@ -2854,8 +3189,10 @@ source = "registry+https://github.com/rust-lang/crates.io-index" "checksum gif 0.10.3 (registry+https://github.com/rust-lang/crates.io-index)" = "471d90201b3b223f3451cd4ad53e34295f16a1df17b1edf3736d47761c3981af" "checksum gimli 0.20.0 (registry+https://github.com/rust-lang/crates.io-index)" = "81dd6190aad0f05ddbbf3245c54ed14ca4aa6dd32f22312b70d8f168c3e3e633" "checksum gl 0.14.0 (registry+https://github.com/rust-lang/crates.io-index)" = "a94edab108827d67608095e269cf862e60d920f144a5026d3dbcfd8b877fb404" +"checksum gl_generator 0.11.0 (registry+https://github.com/rust-lang/crates.io-index)" = "39a23d5e872a275135d66895d954269cf5e8661d234eb1c2480f4ce0d586acbd" "checksum gl_generator 0.13.1 (registry+https://github.com/rust-lang/crates.io-index)" = "ca98bbde17256e02d17336a6bdb5a50f7d0ccacee502e191d3e3d0ec2f96f84a" "checksum gl_generator 0.14.0 (registry+https://github.com/rust-lang/crates.io-index)" = "1a95dfc23a2b4a9a2f5ab41d194f8bfda3cabec42af4e39f08c339eb2a0c124d" +"checksum gleam 0.7.0 (registry+https://github.com/rust-lang/crates.io-index)" = "9ea4f9ba7411ae3f00516401fb811b4f4f37f5c926357f2a033d27f96b74c849" "checksum glutin 0.23.0 (registry+https://github.com/rust-lang/crates.io-index)" = "cf22d4e90c55d9be9f2ad52410e7a2c0d7e9c99d93a13df73a672e7ef4e8c7f7" "checksum glutin_egl_sys 0.1.4 (registry+https://github.com/rust-lang/crates.io-index)" = "772edef3b28b8ad41e4ea202748e65eefe8e5ffd1f4535f1219793dbb20b3d4c" "checksum glutin_emscripten_sys 0.1.1 (registry+https://github.com/rust-lang/crates.io-index)" = "80de4146df76e8a6c32b03007bc764ff3249dcaeb4f675d68a06caf1bac363f1" @@ -2874,6 +3211,7 @@ source = "registry+https://github.com/rust-lang/crates.io-index" "checksum image 0.23.3 (registry+https://github.com/rust-lang/crates.io-index)" = "bfc5483f8d5afd3653b38a196c52294dcb239c3e1a5bade1990353ea13bcf387" "checksum inflate 0.4.5 (registry+https://github.com/rust-lang/crates.io-index)" = "1cdb29978cc5797bd8dcc8e5bf7de604891df2a8dc576973d71a281e916db2ff" "checksum instant 0.1.2 (registry+https://github.com/rust-lang/crates.io-index)" = "6c346c299e3fe8ef94dc10c2c0253d858a69aac1245157a3bf4125915d528caf" +"checksum io-surface 0.12.1 (registry+https://github.com/rust-lang/crates.io-index)" = "2279a6faecd06034f88218f77f7a767693e0957bce0323a96d92747e2760b445" "checksum iovec 0.1.4 (registry+https://github.com/rust-lang/crates.io-index)" = "b2b3ea6ff95e175473f8ffe6a7eb7c00d054240321b84c57051175fe3c1e075e" "checksum itoa 0.4.5 (registry+https://github.com/rust-lang/crates.io-index)" = "b8b7a7c0c47db5545ed3fef7468ee7bb5b74691498139e4b3f6a20685dc6dd8e" "checksum jemalloc-sys 0.3.2 (registry+https://github.com/rust-lang/crates.io-index)" = "0d3b9f3f5c9b31aa0f5ed3260385ac205db665baa41d49bb8338008ae94ede45" @@ -2888,6 +3226,8 @@ source = "registry+https://github.com/rust-lang/crates.io-index" "checksum kurbo 0.5.11 (registry+https://github.com/rust-lang/crates.io-index)" = "bf50e17a1697110c694d47c5b1a6b64faf5eb3ffe5a286df23fb8cd516e33be6" "checksum lazy_static 1.4.0 (registry+https://github.com/rust-lang/crates.io-index)" = "e2abad23fbc42b3700f2f279844dc832adb2b2eb069b2df918f455c4e18cc646" "checksum lazycell 1.2.1 (registry+https://github.com/rust-lang/crates.io-index)" = "b294d6fa9ee409a054354afc4352b0b9ef7ca222c69b8812cbea9e7d2bf3783f" +"checksum leak 0.1.2 (registry+https://github.com/rust-lang/crates.io-index)" = "bd100e01f1154f2908dfa7d02219aeab25d0b9c7fa955164192e3245255a0c73" +"checksum leaky-cow 0.1.1 (registry+https://github.com/rust-lang/crates.io-index)" = "40a8225d44241fd324a8af2806ba635fc7c8a7e9a7de4d5cf3ef54e71f5926fc" "checksum lexical-core 0.6.2 (registry+https://github.com/rust-lang/crates.io-index)" = "d7043aa5c05dd34fb73b47acb8c3708eac428de4545ea3682ed2f11293ebd890" "checksum libc 0.1.12 (registry+https://github.com/rust-lang/crates.io-index)" = "e32a70cf75e5846d53a673923498228bbec6a8624708a9ea5645f075d6276122" "checksum libc 0.2.69 (registry+https://github.com/rust-lang/crates.io-index)" = "99e85c08494b21a9054e7fe1374a732aeadaff3980b6990b94bfd3a70f690005" @@ -2897,6 +3237,7 @@ source = "registry+https://github.com/rust-lang/crates.io-index" "checksum log 0.4.8 (registry+https://github.com/rust-lang/crates.io-index)" = "14b6052be84e6b71ab17edffc2eeabf5c2c3ae1fdb464aae35ac50c67a44e1f7" "checksum lzma-rs 0.1.2 (registry+https://github.com/rust-lang/crates.io-index)" = "ad0606857a51b9088eb75b52d8431b7b7c8656849cc6cb96dde9f3d18a1a4b58" "checksum lzw 0.10.0 (registry+https://github.com/rust-lang/crates.io-index)" = "7d947cbb889ed21c2a84be6ffbaebf5b4e0f4340638cba0444907e38b56be084" +"checksum mach 0.3.2 (registry+https://github.com/rust-lang/crates.io-index)" = "b823e83b2affd8f40a9ee8c29dbc56404c1e34cd2710921f2801e2cf29527afa" "checksum malloc_buf 0.0.6 (registry+https://github.com/rust-lang/crates.io-index)" = "62bb907fe88d54d8d9ce32a3cceab4218ed2f6b7d35617cafe9adf84e43919cb" "checksum matches 0.1.8 (registry+https://github.com/rust-lang/crates.io-index)" = "7ffc5c5338469d4d3ea17d269fa8ea3512ad247247c30bd2df69e68309ed0a08" "checksum maybe-uninit 2.0.0 (registry+https://github.com/rust-lang/crates.io-index)" = "60302e4db3a61da70c0cb7991976248362f30319e88850c487b9b95bbf059e00" @@ -2920,10 +3261,14 @@ source = "registry+https://github.com/rust-lang/crates.io-index" "checksum num-traits 0.2.11 (registry+https://github.com/rust-lang/crates.io-index)" = "c62be47e61d1842b9170f0fdeec8eba98e60e90e5446449a0545e5152acd7096" "checksum num_cpus 1.13.0 (registry+https://github.com/rust-lang/crates.io-index)" = "05499f3756671c15885fee9034446956fff3f243d6077b91e5767df161f766b3" "checksum objc 0.2.7 (registry+https://github.com/rust-lang/crates.io-index)" = "915b1b472bc21c53464d6c8461c9d3af805ba1ef837e1cac254428f4a77177b1" +"checksum objc-foundation 0.1.1 (registry+https://github.com/rust-lang/crates.io-index)" = "1add1b659e36c9607c7aab864a76c7a4c2760cd0cd2e120f3fb8b952c7e22bf9" "checksum objc_exception 0.1.2 (registry+https://github.com/rust-lang/crates.io-index)" = "ad970fb455818ad6cba4c122ad012fae53ae8b4795f86378bce65e4f6bab2ca4" +"checksum objc_id 0.1.1 (registry+https://github.com/rust-lang/crates.io-index)" = "c92d4ddb4bd7b50d730c215ff871754d0da6b2178849f8a2a2ab69712d0c073b" "checksum ordered-float 1.0.2 (registry+https://github.com/rust-lang/crates.io-index)" = "18869315e81473c951eb56ad5558bbc56978562d3ecfb87abb7a1e944cea4518" "checksum osmesa-sys 0.1.2 (registry+https://github.com/rust-lang/crates.io-index)" = "88cfece6e95d2e717e0872a7f53a8684712ad13822a7979bc760b9c77ec0013b" "checksum parking_lot 0.10.2 (registry+https://github.com/rust-lang/crates.io-index)" = "d3a704eb390aafdc107b0e392f56a82b668e3a71366993b5340f5833fd62505e" +"checksum parking_lot 0.9.0 (registry+https://github.com/rust-lang/crates.io-index)" = "f842b1982eb6c2fe34036a4fbfb06dd185a3f5c8edfaacdf7d1ea10b07de6252" +"checksum parking_lot_core 0.6.2 (registry+https://github.com/rust-lang/crates.io-index)" = "b876b1b9e7ac6e1a74a6da34d25c42e17e8862aa409cbbbdcfc8d86c6f3bc62b" "checksum parking_lot_core 0.7.1 (registry+https://github.com/rust-lang/crates.io-index)" = "0e136c1904604defe99ce5fd71a28d473fa60a12255d511aa78a9ddf11237aeb" "checksum percent-encoding 2.1.0 (registry+https://github.com/rust-lang/crates.io-index)" = "d4fd5641d01c8f18a23da7b6fe29298ff4b55afcccdf78973b24cf3175fee32e" "checksum pkg-config 0.3.17 (registry+https://github.com/rust-lang/crates.io-index)" = "05da548ad6865900e60eaba7f589cc0783590a92e940c26953ff81ddbab2d677" @@ -2978,12 +3323,16 @@ source = "registry+https://github.com/rust-lang/crates.io-index" "checksum siphasher 0.2.3 (registry+https://github.com/rust-lang/crates.io-index)" = "0b8de496cf83d4ed58b6be86c3a275b8602f6ffe98d3024a869e124147a9a3ac" "checksum skribo 0.1.0 (registry+https://github.com/rust-lang/crates.io-index)" = "e6e9e713ecb4b6d3047428d060aa46cf4abd1109a961da245359e8f88a529d16" "checksum slab 0.4.2 (registry+https://github.com/rust-lang/crates.io-index)" = "c111b5bd5695e56cffe5129854aa230b39c93a305372fdbb2668ca2394eea9f8" +"checksum smallvec 0.6.13 (registry+https://github.com/rust-lang/crates.io-index)" = "f7b0758c52e15a8b5e3691eae6cc559f08eee9406e548a4477ba4e67770a82b6" "checksum smallvec 1.3.0 (registry+https://github.com/rust-lang/crates.io-index)" = "05720e22615919e4734f6a99ceae50d00226c3c5aca406e102ebc33298214e0a" +"checksum smithay-client-toolkit 0.4.6 (registry+https://github.com/rust-lang/crates.io-index)" = "2ccb8c57049b2a34d2cc2b203fa785020ba0129d31920ef0d317430adaf748fa" "checksum smithay-client-toolkit 0.6.6 (registry+https://github.com/rust-lang/crates.io-index)" = "421c8dc7acf5cb205b88160f8b4cc2c5cfabe210e43b2f80f009f4c1ef910f1d" "checksum stable_deref_trait 1.1.1 (registry+https://github.com/rust-lang/crates.io-index)" = "dba1a27d3efae4351c8051072d619e3ade2820635c3958d826bfea39d59b54c8" "checksum static_assertions 0.3.4 (registry+https://github.com/rust-lang/crates.io-index)" = "7f3eb36b47e512f8f1c9e3d10c2c1965bc992bd9cdb024fa581e2194501c83d3" "checksum stb_truetype 0.3.1 (registry+https://github.com/rust-lang/crates.io-index)" = "f77b6b07e862c66a9f3e62a07588fee67cd90a9135a2b942409f195507b4fb51" "checksum strsim 0.8.0 (registry+https://github.com/rust-lang/crates.io-index)" = "8ea5119cdb4c55b55d432abb513a0429384878c15dde60cc77b1c99de1a95a6a" +"checksum surfman 0.2.0 (git+https://github.com/pcwalton/surfman?rev=bc29e7ae88ca7dd64d3b1c7e185604693290207f)" = "" +"checksum surfman 0.2.0 (registry+https://github.com/rust-lang/crates.io-index)" = "d28b4931b29b2c65fcce61898d6b4dbae2337a60308176c2d5cbff0913740d4a" "checksum svgtypes 0.5.0 (registry+https://github.com/rust-lang/crates.io-index)" = "9c536faaff1a10837cfe373142583f6e27d81e96beba339147e77b67c9f260ff" "checksum swf-fixed 0.1.5 (registry+https://github.com/rust-lang/crates.io-index)" = "6b212c20df50b382c442a4098d7d9f1c94f0b040f0f0a5d120fa3a82fa51e302" "checksum swf-parser 0.10.0 (registry+https://github.com/rust-lang/crates.io-index)" = "847c41c7add3a5a64524fa5883ba4f2ecaccb2e429df9a3d4ed17ad8e379c15f" @@ -2994,6 +3343,7 @@ source = "registry+https://github.com/rust-lang/crates.io-index" "checksum textwrap 0.11.0 (registry+https://github.com/rust-lang/crates.io-index)" = "d326610f408c7a4eb6f51c37c330e496b08506c9457c9d34287ecc38809fb060" "checksum thread_local 1.0.1 (registry+https://github.com/rust-lang/crates.io-index)" = "d40c6d1b69745a6ec6fb1ca717914848da4b44ae29d9b3080cbee91d72a69b14" "checksum tiff 0.4.0 (registry+https://github.com/rust-lang/crates.io-index)" = "002351e428db1eb1d8656d4ca61947c3519ac3191e1c804d4600cd32093b77ad" +"checksum time-point 0.1.1 (registry+https://github.com/rust-lang/crates.io-index)" = "06535c958d6abe68dc4b4ef9e6845f758fc42fe463d0093d0aca40254f03fb14" "checksum toml 0.5.6 (registry+https://github.com/rust-lang/crates.io-index)" = "ffc92d160b1eef40665be3a05630d003936a3bc7da7421277846c2613e92c71a" "checksum ttf-parser 0.3.0 (registry+https://github.com/rust-lang/crates.io-index)" = "a67a691cd15aae8f55fcc6e68efec96ec9e6e3ad967ac16f18681e2268c92037" "checksum unicode-bidi 0.3.4 (registry+https://github.com/rust-lang/crates.io-index)" = "49f2bd0c6468a8230e1db229cff8029217cf623c767ea5d60bfbd42729ea54d5" @@ -3015,11 +3365,17 @@ source = "registry+https://github.com/rust-lang/crates.io-index" "checksum wasm-bindgen-macro 0.2.60 (registry+https://github.com/rust-lang/crates.io-index)" = "8bd151b63e1ea881bb742cd20e1d6127cef28399558f3b5d415289bc41eee3a4" "checksum wasm-bindgen-macro-support 0.2.60 (registry+https://github.com/rust-lang/crates.io-index)" = "d68a5b36eef1be7868f668632863292e37739656a80fc4b9acec7b0bd35a4931" "checksum wasm-bindgen-shared 0.2.60 (registry+https://github.com/rust-lang/crates.io-index)" = "daf76fe7d25ac79748a37538b7daeed1c7a6867c92d3245c12c6222e4a20d639" +"checksum wayland-client 0.21.13 (registry+https://github.com/rust-lang/crates.io-index)" = "49963e5f9eeaf637bfcd1b9f0701c99fd5cd05225eb51035550d4272806f2713" "checksum wayland-client 0.23.6 (registry+https://github.com/rust-lang/crates.io-index)" = "af1080ebe0efabcf12aef2132152f616038f2d7dcbbccf7b2d8c5270fe14bcda" +"checksum wayland-commons 0.21.13 (registry+https://github.com/rust-lang/crates.io-index)" = "40c08896768b667e1df195d88a62a53a2d1351a1ed96188be79c196b35bb32ec" "checksum wayland-commons 0.23.6 (registry+https://github.com/rust-lang/crates.io-index)" = "bb66b0d1a27c39bbce712b6372131c6e25149f03ffb0cd017cf8f7de8d66dbdb" +"checksum wayland-protocols 0.21.13 (registry+https://github.com/rust-lang/crates.io-index)" = "4afde2ea2a428eee6d7d2c8584fdbe8b82eee8b6c353e129a434cd6e07f42145" "checksum wayland-protocols 0.23.6 (registry+https://github.com/rust-lang/crates.io-index)" = "6cc286643656742777d55dc8e70d144fa4699e426ca8e9d4ef454f4bf15ffcf9" +"checksum wayland-scanner 0.21.13 (registry+https://github.com/rust-lang/crates.io-index)" = "bf3828c568714507315ee425a9529edc4a4aa9901409e373e9e0027e7622b79e" "checksum wayland-scanner 0.23.6 (registry+https://github.com/rust-lang/crates.io-index)" = "93b02247366f395b9258054f964fe293ddd019c3237afba9be2ccbe9e1651c3d" +"checksum wayland-sys 0.21.13 (registry+https://github.com/rust-lang/crates.io-index)" = "520ab0fd578017a0ee2206623ba9ef4afe5e8f23ca7b42f6acfba2f4e66b1628" "checksum wayland-sys 0.23.6 (registry+https://github.com/rust-lang/crates.io-index)" = "d94e89a86e6d6d7c7c9b19ebf48a03afaac4af6bc22ae570e9a24124b75358f4" +"checksum wayland-sys 0.24.1 (registry+https://github.com/rust-lang/crates.io-index)" = "537500923d50be11d95a63c4cb538145e4c82edf61296b7debc1f94a1a6514ed" "checksum web-sys 0.3.37 (registry+https://github.com/rust-lang/crates.io-index)" = "2d6f51648d8c56c366144378a33290049eafdd784071077f6fe37dae64c1c4cb" "checksum winapi 0.2.8 (registry+https://github.com/rust-lang/crates.io-index)" = "167dc9d6949a9b857f3451275e911c3f44255842c1f7a76f33c55103a909087a" "checksum winapi 0.3.8 (registry+https://github.com/rust-lang/crates.io-index)" = "8093091eeb260906a183e6ae1abdba2ef5ef2257a21801128899c3fc699229c6" @@ -3027,9 +3383,11 @@ source = "registry+https://github.com/rust-lang/crates.io-index" "checksum winapi-i686-pc-windows-gnu 0.4.0 (registry+https://github.com/rust-lang/crates.io-index)" = "ac3b87c63620426dd9b991e5ce0329eff545bccbbb34f3be09ff6fb6ab51b7b6" "checksum winapi-util 0.1.4 (registry+https://github.com/rust-lang/crates.io-index)" = "fa515c5163a99cc82bab70fd3bfdd36d827be85de63737b40fcef2ce084a436e" "checksum winapi-x86_64-pc-windows-gnu 0.4.0 (registry+https://github.com/rust-lang/crates.io-index)" = "712e227841d057c1ee1cd2fb22fa7e5a5461ae8e48fa2ca79ec42cfc1931183f" +"checksum winit 0.19.3 (registry+https://github.com/rust-lang/crates.io-index)" = "7d0da905e61ae52d55c5ca6f8bea1e09daf5e325b6c77b0947c65a5179b49f5f" "checksum winit 0.21.0 (registry+https://github.com/rust-lang/crates.io-index)" = "65a5c1a5ef76ac31cc97ad29489acdbed2178f3fc12ca00ee6cb11d60adb5a3a" "checksum wio 0.2.2 (registry+https://github.com/rust-lang/crates.io-index)" = "5d129932f4644ac2396cb456385cbf9e63b5b30c6e8dc4820bdca4eb082037a5" "checksum ws2_32-sys 0.2.1 (registry+https://github.com/rust-lang/crates.io-index)" = "d59cefebd0c892fa2dd6de581e937301d8552cb44489cdff035c6187cb63fa5e" +"checksum x11 2.18.2 (registry+https://github.com/rust-lang/crates.io-index)" = "77ecd092546cb16f25783a5451538e73afc8d32e242648d54f4ae5459ba1e773" "checksum x11-dl 2.18.5 (registry+https://github.com/rust-lang/crates.io-index)" = "2bf981e3a5b3301209754218f962052d4d9ee97e478f4d26d4a6eced34c1fef8" "checksum xdg 2.2.0 (registry+https://github.com/rust-lang/crates.io-index)" = "d089681aa106a86fade1b0128fb5daf07d5867a509ab036d99988dec80429a57" "checksum xml-rs 0.8.2 (registry+https://github.com/rust-lang/crates.io-index)" = "2bb76e5c421bbbeb8924c60c030331b345555024d56261dae8f3e786ed817c23" diff --git a/c/src/lib.rs b/c/src/lib.rs index 369e326e..ecf4f6c7 100644 --- a/c/src/lib.rs +++ b/c/src/lib.rs @@ -39,7 +39,7 @@ use std::slice; use std::str; #[cfg(all(target_os = "macos", not(feature = "pf-gl")))] -use metal::{CAMetalLayer, CoreAnimationLayerRef}; +use metal::{CAMetalLayer, CoreAnimationLayerRef, Device}; #[cfg(all(target_os = "macos", not(feature = "pf-gl")))] use pathfinder_metal::MetalDevice; @@ -617,7 +617,9 @@ pub unsafe extern "C" fn PFSceneProxyBuildAndRenderMetal(scene_proxy: PFScenePro #[no_mangle] pub unsafe extern "C" fn PFMetalDeviceCreate(layer: *mut CAMetalLayer) -> PFMetalDeviceRef { - Box::into_raw(Box::new(MetalDevice::new(CoreAnimationLayerRef::from_ptr(layer)))) + let device = Device::system_default().expect("Failed to get Metal system default device!"); + let layer = CoreAnimationLayerRef::from_ptr(layer); + Box::into_raw(Box::new(MetalDevice::new(device, layer.next_drawable().unwrap()))) } #[cfg(all(target_os = "macos", not(feature = "pf-gl")))] @@ -626,12 +628,6 @@ pub unsafe extern "C" fn PFMetalDeviceDestroy(device: PFMetalDeviceRef) { drop(Box::from_raw(device)) } -#[cfg(all(target_os = "macos", not(feature = "pf-gl")))] -#[no_mangle] -pub unsafe extern "C" fn PFMetalDevicePresentDrawable(device: PFMetalDeviceRef) { - (*device).present_drawable() -} - // `renderer` #[no_mangle] @@ -806,6 +802,8 @@ impl PFRendererOptions { } else { None }, + // TODO(pcwalton): Expose this in the C API. + use_compute: false, } } } diff --git a/demo/common/Cargo.toml b/demo/common/Cargo.toml index ff0e8564..169fd9df 100644 --- a/demo/common/Cargo.toml +++ b/demo/common/Cargo.toml @@ -57,5 +57,8 @@ path = "../../ui" [target.'cfg(target_os = "macos")'.dependencies] metal = "0.17" +[target.'cfg(target_os = "macos")'.dependencies.io-surface] +version = "0.12" + [target.'cfg(target_os = "macos")'.dependencies.pathfinder_metal] path = "../../metal" diff --git a/demo/common/src/device.rs b/demo/common/src/device.rs index 6d685b02..f477b4b0 100644 --- a/demo/common/src/device.rs +++ b/demo/common/src/device.rs @@ -29,7 +29,7 @@ where D: Device, { pub fn new(device: &D, resources: &dyn ResourceLoader) -> GroundProgram { - let program = device.create_program(resources, "demo_ground"); + let program = device.create_raster_program(resources, "demo_ground"); let transform_uniform = device.get_uniform(&program, "Transform"); let gridline_count_uniform = device.get_uniform(&program, "GridlineCount"); let ground_color_uniform = device.get_uniform(&program, "GroundColor"); diff --git a/demo/common/src/lib.rs b/demo/common/src/lib.rs index 9d6c35b0..3f1a24a0 100644 --- a/demo/common/src/lib.rs +++ b/demo/common/src/lib.rs @@ -116,13 +116,13 @@ pub struct DemoApp where W: Window { } impl DemoApp where W: Window { - pub fn new(window: W, window_size: WindowSize, mut options: Options) -> DemoApp { + pub fn new(window: W, window_size: WindowSize, options: Options) -> DemoApp { let expire_message_event_id = window.create_user_event_id(); let device; #[cfg(all(target_os = "macos", not(feature = "pf-gl")))] - { - device = DeviceImpl::new(window.metal_layer()); + unsafe { + device = DeviceImpl::new(window.metal_device(), window.metal_io_surface()); } #[cfg(any(not(target_os = "macos"), feature = "pf-gl"))] { @@ -131,14 +131,14 @@ impl DemoApp where W: Window { let resources = window.resource_loader(); - // Read command line options. - options.command_line_overrides(); - // Set up the executor. let executor = DemoExecutor::new(options.jobs); let mut ui_model = DemoUIModel::new(&options); - let render_options = RendererOptions { background_color: None }; + let render_options = RendererOptions { + background_color: None, + use_compute: options.compute, + }; let filter = build_filter(&ui_model); @@ -624,6 +624,8 @@ pub struct Options { pub input_path: SVGPath, pub ui: UIVisibility, pub background_color: BackgroundColor, + pub high_performance_gpu: bool, + pub compute: bool, hidden_field_for_future_proofing: (), } @@ -635,13 +637,15 @@ impl Default for Options { input_path: SVGPath::Default, ui: UIVisibility::All, background_color: BackgroundColor::Light, + high_performance_gpu: false, + compute: false, hidden_field_for_future_proofing: (), } } } impl Options { - fn command_line_overrides(&mut self) { + pub fn command_line_overrides(&mut self) { let matches = App::new("tile-svg") .arg( Arg::with_name("jobs") @@ -681,6 +685,18 @@ impl Options { .possible_values(&["light", "dark", "transparent"]) .help("The background color to use"), ) + .arg( + Arg::with_name("high-performance-gpu") + .short("g") + .long("high-performance-gpu") + .help("Use the high-performance (discrete) GPU, if available") + ) + .arg( + Arg::with_name("compute") + .short("c") + .long("compute") + .help("Use compute shaders for certain tasks, if available") + ) .arg( Arg::with_name("INPUT") .help("Path to the SVG file to render") @@ -714,6 +730,14 @@ impl Options { }; } + if matches.is_present("high-performance-gpu") { + self.high_performance_gpu = true; + } + + if matches.is_present("compute") { + self.compute = true; + } + if let Some(path) = matches.value_of("INPUT") { self.input_path = SVGPath::Path(PathBuf::from(path)); }; diff --git a/demo/common/src/renderer.rs b/demo/common/src/renderer.rs index 71f9c89e..a142cd99 100644 --- a/demo/common/src/renderer.rs +++ b/demo/common/src/renderer.rs @@ -89,7 +89,10 @@ impl DemoApp where W: Window { Mode::ThreeD => None, Mode::VR => Some(ColorF::transparent_black()), }; - self.renderer.set_options(RendererOptions { background_color: clear_color }); + self.renderer.set_options(RendererOptions { + background_color: clear_color, + use_compute: self.options.compute, + }); scene_count } @@ -229,6 +232,7 @@ impl DemoApp where W: Window { vertex_array: &self.ground_vertex_array.vertex_array, primitive: Primitive::Triangles, textures: &[], + images: &[], uniforms: &[ (&self.ground_program.transform_uniform, UniformData::from_transform_3d(&transform)), diff --git a/demo/common/src/window.rs b/demo/common/src/window.rs index 2fd45ebd..bede4502 100644 --- a/demo/common/src/window.rs +++ b/demo/common/src/window.rs @@ -18,9 +18,11 @@ use rayon::ThreadPoolBuilder; use std::path::PathBuf; #[cfg(all(target_os = "macos", not(feature = "pf-gl")))] -use metal::CoreAnimationLayerRef; +use io_surface::IOSurfaceRef; #[cfg(all(target_os = "macos", not(feature = "pf-gl")))] -use pathfinder_metal::MetalDevice; +use metal::Device as MetalDevice; +#[cfg(all(target_os = "macos", not(feature = "pf-gl")))] +use pathfinder_metal::MetalDevice as PathfinderMetalDevice; #[cfg(any(not(target_os = "macos"), feature = "pf-gl"))] use gl::types::GLuint; @@ -36,9 +38,11 @@ pub trait Window { fn present(&mut self, device: &mut GLDevice); #[cfg(all(target_os = "macos", not(feature = "pf-gl")))] - fn metal_layer(&self) -> &CoreAnimationLayerRef; + fn metal_device(&self) -> MetalDevice; #[cfg(all(target_os = "macos", not(feature = "pf-gl")))] - fn present(&mut self, device: &mut MetalDevice); + fn metal_io_surface(&self) -> IOSurfaceRef; + #[cfg(all(target_os = "macos", not(feature = "pf-gl")))] + fn present(&mut self, device: &mut PathfinderMetalDevice); fn make_current(&mut self, view: View); fn viewport(&self, view: View) -> RectI; diff --git a/demo/native/Cargo.toml b/demo/native/Cargo.toml index 0d6d406f..529bfd5f 100644 --- a/demo/native/Cargo.toml +++ b/demo/native/Cargo.toml @@ -11,10 +11,13 @@ pf-no-simd = ["pathfinder_simd/pf-no-simd"] [dependencies] color-backtrace = "0.3" gl = "0.14" +lazy_static = "1" nfd = "0.0.4" pretty_env_logger = "0.4" -sdl2 = "0.33" -sdl2-sys = "0.33" + +[dependencies.euclid] +version = "0.20" +features = [] [dependencies.pathfinder_demo] path = "../common" @@ -34,9 +37,19 @@ path = "../../resources" [dependencies.pathfinder_simd] path = "../../simd" +[dependencies.surfman] +# version = "0.2" +path = "/Users/pcwalton/Source/surfman/surfman" +features = ["sm-winit"] + +[dependencies.winit] +version = "<0.19.4" # 0.19.4 causes build errors https://github.com/rust-windowing/winit/pull/1105 + [target.'cfg(target_os = "macos")'.dependencies] foreign-types = "0.3" +io-surface = "0.12" metal = "0.17" +objc = "0.2" [target.'cfg(target_os = "macos")'.dependencies.pathfinder_metal] path = "../../metal" diff --git a/demo/native/src/main.rs b/demo/native/src/main.rs index 96f9a40e..f42a81e2 100644 --- a/demo/native/src/main.rs +++ b/demo/native/src/main.rs @@ -10,38 +10,48 @@ //! A demo app for Pathfinder using SDL 2. +#[macro_use] +extern crate lazy_static; + +#[cfg(all(target_os = "macos", not(feature = "pf-gl")))] +extern crate objc; + +use euclid::default::Size2D; use nfd::Response; use pathfinder_demo::window::{Event, Keycode, SVGPath, View, Window, WindowSize}; use pathfinder_demo::{DemoApp, Options}; use pathfinder_geometry::rect::RectI; -use pathfinder_geometry::vector::vec2i; +use pathfinder_geometry::vector::{Vector2I, vec2i}; use pathfinder_resources::ResourceLoader; use pathfinder_resources::fs::FilesystemResourceLoader; -use sdl2::event::{Event as SDLEvent, WindowEvent}; -use sdl2::keyboard::Keycode as SDLKeycode; -use sdl2::video::Window as SDLWindow; -use sdl2::{EventPump, EventSubsystem, Sdl, VideoSubsystem}; -use sdl2_sys::{SDL_Event, SDL_UserEvent}; +use std::cell::Cell; +use std::collections::VecDeque; use std::path::PathBuf; -use std::ptr; +use std::sync::Mutex; +use surfman::{SurfaceAccess, SurfaceType, declare_surfman}; +use winit::{ControlFlow, ElementState, Event as WinitEvent, EventsLoop, EventsLoopProxy}; +use winit::{MouseButton, VirtualKeyCode, Window as WinitWindow, WindowBuilder, WindowEvent}; +use winit::dpi::LogicalSize; +#[cfg(any(not(target_os = "macos"), feature = "pf-gl"))] +use gl::types::GLuint; +#[cfg(any(not(target_os = "macos"), feature = "pf-gl"))] +use gl; +#[cfg(any(not(target_os = "macos"), feature = "pf-gl"))] +use surfman::{Connection, Context, ContextAttributeFlags, ContextAttributes}; +#[cfg(any(not(target_os = "macos"), feature = "pf-gl"))] +use surfman::{Device, GLVersion as SurfmanGLVersion}; #[cfg(all(target_os = "macos", not(feature = "pf-gl")))] -use foreign_types::ForeignTypeRef; -#[cfg(all(target_os = "macos", not(feature = "pf-gl")))] -use metal::{CAMetalLayer, CoreAnimationLayerRef}; +use io_surface::IOSurfaceRef; #[cfg(all(target_os = "macos", not(feature = "pf-gl")))] use pathfinder_metal::MetalDevice; #[cfg(all(target_os = "macos", not(feature = "pf-gl")))] -use sdl2::hint; -#[cfg(all(target_os = "macos", not(feature = "pf-gl")))] -use sdl2::render::Canvas; -#[cfg(all(target_os = "macos", not(feature = "pf-gl")))] -use sdl2_sys::SDL_RenderGetMetalLayer; +use surfman::{NativeDevice, SystemConnection, SystemDevice, SystemSurface}; + +declare_surfman!(); #[cfg(any(not(target_os = "macos"), feature = "pf-gl"))] use pathfinder_gl::{GLDevice, GLVersion}; -#[cfg(any(not(target_os = "macos"), feature = "pf-gl"))] -use sdl2::video::{GLContext, GLProfile}; #[cfg(not(windows))] use jemallocator; @@ -53,13 +63,21 @@ static ALLOC: jemallocator::Jemalloc = jemallocator::Jemalloc; const DEFAULT_WINDOW_WIDTH: u32 = 1067; const DEFAULT_WINDOW_HEIGHT: u32 = 800; +lazy_static! { + static ref EVENT_QUEUE: Mutex> = Mutex::new(None); +} + fn main() { color_backtrace::install(); pretty_env_logger::init(); - let window = WindowImpl::new(); + // Read command line options. + let mut options = Options::default(); + options.command_line_overrides(); + + let window = WindowImpl::new(&options); let window_size = window.size(); - let options = Options::default(); + let mut app = DemoApp::new(window, window_size, options); while !app.should_exit { @@ -72,6 +90,7 @@ fn main() { } let scene_count = app.prepare_frame(events); + app.draw_scene(); app.begin_compositing(); for scene_index in 0..scene_count { @@ -81,28 +100,46 @@ fn main() { } } -thread_local! { - static SDL_CONTEXT: Sdl = sdl2::init().unwrap(); - static SDL_VIDEO: VideoSubsystem = SDL_CONTEXT.with(|context| context.video().unwrap()); - static SDL_EVENT: EventSubsystem = SDL_CONTEXT.with(|context| context.event().unwrap()); -} - struct WindowImpl { + window: WinitWindow, + #[cfg(any(not(target_os = "macos"), feature = "pf-gl"))] - window: SDLWindow, + context: Context, #[cfg(any(not(target_os = "macos"), feature = "pf-gl"))] - gl_context: GLContext, + #[allow(dead_code)] + connection: Connection, + #[cfg(any(not(target_os = "macos"), feature = "pf-gl"))] + device: Device, #[cfg(all(target_os = "macos", not(feature = "pf-gl")))] - canvas: Canvas, + #[allow(dead_code)] + connection: SystemConnection, #[cfg(all(target_os = "macos", not(feature = "pf-gl")))] - metal_layer: *mut CAMetalLayer, + device: SystemDevice, + #[cfg(all(target_os = "macos", not(feature = "pf-gl")))] + metal_device: NativeDevice, + #[cfg(all(target_os = "macos", not(feature = "pf-gl")))] + surface: SystemSurface, + + event_loop: EventsLoop, + pending_events: VecDeque, + mouse_position: Vector2I, + mouse_down: bool, + next_user_event_id: Cell, - event_pump: EventPump, #[allow(dead_code)] resource_loader: FilesystemResourceLoader, - selected_file: Option, - open_svg_message_type: u32, +} + +struct EventQueue { + event_loop_proxy: EventsLoopProxy, + pending_custom_events: VecDeque, +} + +#[derive(Clone)] +enum CustomEvent { + User { message_type: u32, message_data: u32 }, + OpenSVG(PathBuf), } impl Window for WindowImpl { @@ -111,26 +148,35 @@ impl Window for WindowImpl { GLVersion::GL3 } + #[cfg(any(not(target_os = "macos"), feature = "pf-gl"))] + fn gl_default_framebuffer(&self) -> GLuint { + self.device.context_surface_info(&self.context).unwrap().unwrap().framebuffer_object + } + #[cfg(all(target_os = "macos", not(feature = "pf-gl")))] - fn metal_layer(&self) -> &CoreAnimationLayerRef { - unsafe { CoreAnimationLayerRef::from_ptr(self.metal_layer) } + fn metal_device(&self) -> metal::Device { + self.metal_device.0.clone() + } + + #[cfg(all(target_os = "macos", not(feature = "pf-gl")))] + fn metal_io_surface(&self) -> IOSurfaceRef { + self.device.native_surface(&self.surface).0 } fn viewport(&self, view: View) -> RectI { - let (width, height) = self.window().drawable_size(); - let mut width = width as i32; - let height = height as i32; + let WindowSize { logical_size, backing_scale_factor } = self.size(); + let mut size = (logical_size.to_f32() * backing_scale_factor).to_i32(); let mut x_offset = 0; if let View::Stereo(index) = view { - width = width / 2; - x_offset = width * (index as i32); + size.set_x(size.x() / 2); + x_offset = size.x() * (index as i32); } - RectI::new(vec2i(x_offset, 0), vec2i(width, height)) + RectI::new(vec2i(x_offset, 0), size) } #[cfg(any(not(target_os = "macos"), feature = "pf-gl"))] fn make_current(&mut self, _view: View) { - self.window().gl_make_current(&self.gl_context).unwrap(); + self.device.make_context_current(&self.context).unwrap(); } #[cfg(all(target_os = "macos", not(feature = "pf-gl")))] @@ -138,40 +184,30 @@ impl Window for WindowImpl { #[cfg(any(not(target_os = "macos"), feature = "pf-gl"))] fn present(&mut self, _: &mut GLDevice) { - self.window().gl_swap_window(); + let mut surface = self.device + .unbind_surface_from_context(&mut self.context) + .unwrap() + .unwrap(); + self.device.present_surface(&mut self.context, &mut surface).unwrap(); + self.device.bind_surface_to_context(&mut self.context, surface).unwrap(); } #[cfg(all(target_os = "macos", not(feature = "pf-gl")))] - fn present(&mut self, device: &mut MetalDevice) { - device.present_drawable(); + fn present(&mut self, metal_device: &mut MetalDevice) { + self.device.present_surface(&mut self.surface).expect("Failed to present surface!"); + metal_device.swap_texture(self.device.native_surface(&self.surface).0); } fn resource_loader(&self) -> &dyn ResourceLoader { &self.resource_loader } - fn create_user_event_id(&self) -> u32 { - SDL_EVENT.with(|sdl_event| unsafe { sdl_event.register_event().unwrap() }) - } - - fn push_user_event(message_type: u32, message_data: u32) { - unsafe { - let mut user_event = SDL_UserEvent { - timestamp: 0, - windowID: 0, - type_: message_type, - code: message_data as i32, - data1: ptr::null_mut(), - data2: ptr::null_mut(), - }; - sdl2_sys::SDL_PushEvent(&mut user_event as *mut SDL_UserEvent as *mut SDL_Event); - } - } - fn present_open_svg_dialog(&mut self) { if let Ok(Response::Okay(path)) = nfd::open_file_dialog(Some("svg"), None) { - self.selected_file = Some(PathBuf::from(path)); - WindowImpl::push_user_event(self.open_svg_message_type, 0); + let mut event_queue = EVENT_QUEUE.lock().unwrap(); + let event_queue = event_queue.as_mut().unwrap(); + event_queue.pending_custom_events.push_back(CustomEvent::OpenSVG(PathBuf::from(path))); + drop(event_queue.event_loop_proxy.wakeup()); } } @@ -181,181 +217,272 @@ impl Window for WindowImpl { _ => Err(()), } } + + fn create_user_event_id(&self) -> u32 { + let id = self.next_user_event_id.get(); + self.next_user_event_id.set(id + 1); + id + } + + fn push_user_event(message_type: u32, message_data: u32) { + let mut event_queue = EVENT_QUEUE.lock().unwrap(); + let event_queue = event_queue.as_mut().unwrap(); + event_queue.pending_custom_events.push_back(CustomEvent::User { + message_type, + message_data, + }); + drop(event_queue.event_loop_proxy.wakeup()); + } } impl WindowImpl { #[cfg(any(not(target_os = "macos"), feature = "pf-gl"))] - fn new() -> WindowImpl { - SDL_VIDEO.with(|sdl_video| { - SDL_EVENT.with(|sdl_event| { - let (window, gl_context, event_pump); + fn new(options: &Options) -> WindowImpl { + let event_loop = EventsLoop::new(); + let window_size = Size2D::new(DEFAULT_WINDOW_WIDTH, DEFAULT_WINDOW_HEIGHT); + let logical_size = LogicalSize::new(window_size.width as f64, window_size.height as f64); + let window = WindowBuilder::new().with_title("Pathfinder Demo") + .with_dimensions(logical_size) + .build(&event_loop) + .unwrap(); + window.show(); - let gl_attributes = sdl_video.gl_attr(); - gl_attributes.set_context_profile(GLProfile::Core); - gl_attributes.set_context_version(3, 3); - gl_attributes.set_depth_size(24); - gl_attributes.set_stencil_size(8); + let connection = Connection::from_winit_window(&window).unwrap(); + let native_widget = connection.create_native_widget_from_winit_window(&window).unwrap(); - window = sdl_video - .window( - "Pathfinder Demo", - DEFAULT_WINDOW_WIDTH, - DEFAULT_WINDOW_HEIGHT, - ) - .opengl() - .resizable() - .allow_highdpi() - .build() - .unwrap(); + let adapter = if options.high_performance_gpu { + connection.create_hardware_adapter().unwrap() + } else { + connection.create_low_power_adapter().unwrap() + }; - gl_context = window.gl_create_context().unwrap(); - gl::load_with(|name| sdl_video.gl_get_proc_address(name) as *const _); + let mut device = connection.create_device(&adapter).unwrap(); - event_pump = SDL_CONTEXT.with(|sdl_context| sdl_context.event_pump().unwrap()); + let context_attributes = ContextAttributes { + version: SurfmanGLVersion::new(3, 0), + flags: ContextAttributeFlags::ALPHA, + }; + let context_descriptor = device.create_context_descriptor(&context_attributes).unwrap(); - let resource_loader = FilesystemResourceLoader::locate(); + let surface_type = SurfaceType::Widget { native_widget }; + let mut context = device.create_context(&context_descriptor).unwrap(); + let surface = device.create_surface(&context, SurfaceAccess::GPUOnly, surface_type) + .unwrap(); + device.bind_surface_to_context(&mut context, surface).unwrap(); + device.make_context_current(&context).unwrap(); - let open_svg_message_type = unsafe { sdl_event.register_event().unwrap() }; + gl::load_with(|symbol_name| device.get_proc_address(&context, symbol_name)); - WindowImpl { - window, - event_pump, - gl_context, - resource_loader, - open_svg_message_type, - selected_file: None, - } - }) - }) + let resource_loader = FilesystemResourceLoader::locate(); + + *EVENT_QUEUE.lock().unwrap() = Some(EventQueue { + event_loop_proxy: event_loop.create_proxy(), + pending_custom_events: VecDeque::new(), + }); + + WindowImpl { + window, + event_loop, + connection, + context, + device, + next_user_event_id: Cell::new(0), + pending_events: VecDeque::new(), + mouse_position: vec2i(0, 0), + mouse_down: false, + resource_loader, + } } #[cfg(all(target_os = "macos", not(feature = "pf-gl")))] - fn new() -> WindowImpl { - assert!(hint::set("SDL_RENDER_DRIVER", "metal")); + fn new(options: &Options) -> WindowImpl { + let event_loop = EventsLoop::new(); + let window_size = Size2D::new(DEFAULT_WINDOW_WIDTH, DEFAULT_WINDOW_HEIGHT); + let logical_size = LogicalSize::new(window_size.width as f64, window_size.height as f64); + let window = WindowBuilder::new().with_title("Pathfinder Demo") + .with_dimensions(logical_size) + .build(&event_loop) + .unwrap(); + window.show(); - SDL_VIDEO.with(|sdl_video| { - SDL_EVENT.with(|sdl_event| { - let window = sdl_video - .window( - "Pathfinder Demo", - DEFAULT_WINDOW_WIDTH, - DEFAULT_WINDOW_HEIGHT, - ) - .opengl() - .resizable() - .allow_highdpi() - .build() - .unwrap(); + let connection = SystemConnection::from_winit_window(&window).unwrap(); + let native_widget = connection.create_native_widget_from_winit_window(&window).unwrap(); - let canvas = window.into_canvas().present_vsync().build().unwrap(); - let metal_layer = unsafe { - SDL_RenderGetMetalLayer(canvas.raw()) as *mut CAMetalLayer - }; + let adapter = if options.high_performance_gpu { + connection.create_hardware_adapter().unwrap() + } else { + connection.create_low_power_adapter().unwrap() + }; - let event_pump = SDL_CONTEXT.with(|sdl_context| sdl_context.event_pump().unwrap()); + let mut device = connection.create_device(&adapter).unwrap(); + let native_device = device.native_device(); - let resource_loader = FilesystemResourceLoader::locate(); + let surface_type = SurfaceType::Widget { native_widget }; + let surface = device.create_surface(SurfaceAccess::GPUOnly, surface_type).unwrap(); - let open_svg_message_type = unsafe { sdl_event.register_event().unwrap() }; + let resource_loader = FilesystemResourceLoader::locate(); - WindowImpl { - event_pump, - canvas, - metal_layer, - resource_loader, - open_svg_message_type, - selected_file: None, - } - }) - }) + *EVENT_QUEUE.lock().unwrap() = Some(EventQueue { + event_loop_proxy: event_loop.create_proxy(), + pending_custom_events: VecDeque::new(), + }); + + WindowImpl { + window, + event_loop, + connection, + device, + metal_device: native_device, + surface, + next_user_event_id: Cell::new(0), + pending_events: VecDeque::new(), + mouse_position: vec2i(0, 0), + mouse_down: false, + resource_loader, + } } - #[cfg(any(not(target_os = "macos"), feature = "pf-gl"))] - fn window(&self) -> &SDLWindow { &self.window } - #[cfg(all(target_os = "macos", not(feature = "pf-gl")))] - fn window(&self) -> &SDLWindow { self.canvas.window() } + fn window(&self) -> &WinitWindow { &self.window } fn size(&self) -> WindowSize { - let (logical_width, logical_height) = self.window().size(); - let (drawable_width, _) = self.window().drawable_size(); + let window = self.window(); + let (monitor, size) = (window.get_current_monitor(), window.get_inner_size().unwrap()); + WindowSize { - logical_size: vec2i(logical_width as i32, logical_height as i32), - backing_scale_factor: drawable_width as f32 / logical_width as f32, + logical_size: vec2i(size.width as i32, size.height as i32), + backing_scale_factor: monitor.get_hidpi_factor() as f32, } } fn get_event(&mut self) -> Event { - loop { - let sdl_event = self.event_pump.wait_event(); - if let Some(event) = self.convert_sdl_event(sdl_event) { - return event; - } + if self.pending_events.is_empty() { + let window = &self.window; + let mouse_position = &mut self.mouse_position; + let mouse_down = &mut self.mouse_down; + let pending_events = &mut self.pending_events; + self.event_loop.run_forever(|winit_event| { + //println!("blocking {:?}", winit_event); + match convert_winit_event(winit_event, + window, + mouse_position, + mouse_down) { + Some(event) => { + //println!("handled"); + pending_events.push_back(event); + ControlFlow::Break + } + None => { + ControlFlow::Continue + } + } + }); } + + self.pending_events.pop_front().expect("Where's the event?") } fn try_get_event(&mut self) -> Option { - loop { - let sdl_event = self.event_pump.poll_event()?; - if let Some(event) = self.convert_sdl_event(sdl_event) { - return Some(event); - } - } - } - - fn convert_sdl_event(&self, sdl_event: SDLEvent) -> Option { - match sdl_event { - SDLEvent::User { type_, .. } if type_ == self.open_svg_message_type => Some( - Event::OpenSVG(SVGPath::Path(self.selected_file.clone().unwrap())), - ), - SDLEvent::User { type_, code, .. } => Some(Event::User { - message_type: type_, - message_data: code as u32, - }), - SDLEvent::MouseButtonDown { x, y, .. } => Some(Event::MouseDown(vec2i(x, y))), - SDLEvent::MouseMotion { - x, y, mousestate, .. - } => { - let position = vec2i(x, y); - if mousestate.left() { - Some(Event::MouseDragged(position)) - } else { - Some(Event::MouseMoved(position)) + if self.pending_events.is_empty() { + let window = &self.window; + let mouse_position = &mut self.mouse_position; + let mouse_down = &mut self.mouse_down; + let pending_events = &mut self.pending_events; + self.event_loop.poll_events(|winit_event| { + //println!("nonblocking {:?}", winit_event); + if let Some(event) = convert_winit_event(winit_event, + window, + mouse_position, + mouse_down) { + //println!("handled"); + pending_events.push_back(event); } - } - SDLEvent::Quit { .. } => Some(Event::Quit), - SDLEvent::Window { - win_event: WindowEvent::SizeChanged(..), - .. - } => Some(Event::WindowResized(self.size())), - SDLEvent::KeyDown { - keycode: Some(sdl_keycode), - .. - } => self.convert_sdl_keycode(sdl_keycode).map(Event::KeyDown), - SDLEvent::KeyUp { - keycode: Some(sdl_keycode), - .. - } => self.convert_sdl_keycode(sdl_keycode).map(Event::KeyUp), - SDLEvent::MultiGesture { d_dist, .. } => { - let mouse_state = self.event_pump.mouse_state(); - let center = vec2i(mouse_state.x(), mouse_state.y()); - Some(Event::Zoom(d_dist, center)) - } - _ => None, - } - } - - fn convert_sdl_keycode(&self, sdl_keycode: SDLKeycode) -> Option { - match sdl_keycode { - SDLKeycode::Escape => Some(Keycode::Escape), - SDLKeycode::Tab => Some(Keycode::Tab), - sdl_keycode - if sdl_keycode as i32 >= SDLKeycode::A as i32 - && sdl_keycode as i32 <= SDLKeycode::Z as i32 => - { - let offset = (sdl_keycode as i32 - SDLKeycode::A as i32) as u8; - Some(Keycode::Alphanumeric(offset + b'a')) - } - _ => None, + }); } + self.pending_events.pop_front() } } + +fn convert_winit_event(winit_event: WinitEvent, + window: &WinitWindow, + mouse_position: &mut Vector2I, + mouse_down: &mut bool) + -> Option { + match winit_event { + WinitEvent::Awakened => { + let mut event_queue = EVENT_QUEUE.lock().unwrap(); + let event_queue = event_queue.as_mut().unwrap(); + match event_queue.pending_custom_events + .pop_front() + .expect("`Awakened` with no pending custom event!") { + CustomEvent::OpenSVG(svg_path) => Some(Event::OpenSVG(SVGPath::Path(svg_path))), + CustomEvent::User { message_data, message_type } => { + Some(Event::User { message_data, message_type }) + } + } + } + WinitEvent::WindowEvent { event: window_event, .. } => { + match window_event { + WindowEvent::MouseInput { + state: ElementState::Pressed, + button: MouseButton::Left, + .. + } => { + *mouse_down = true; + Some(Event::MouseDown(*mouse_position)) + } + WindowEvent::MouseInput { + state: ElementState::Released, + button: MouseButton::Left, + .. + } => { + *mouse_down = false; + None + } + WindowEvent::CursorMoved { position, .. } => { + *mouse_position = vec2i(position.x as i32, position.y as i32); + if *mouse_down { + Some(Event::MouseDragged(*mouse_position)) + } else { + Some(Event::MouseMoved(*mouse_position)) + } + } + WindowEvent::KeyboardInput { input, .. } => { + input.virtual_keycode.and_then(|virtual_keycode| { + match virtual_keycode { + VirtualKeyCode::Escape => Some(Keycode::Escape), + VirtualKeyCode::Tab => Some(Keycode::Tab), + virtual_keycode => { + let vk = virtual_keycode as u32; + let vk_a = VirtualKeyCode::A as u32; + let vk_z = VirtualKeyCode::Z as u32; + if vk >= vk_a && vk <= vk_z { + let character = ((vk - vk_a) + 'A' as u32) as u8; + Some(Keycode::Alphanumeric(character)) + } else { + None + } + } + } + }).map(|keycode| { + match input.state { + ElementState::Pressed => Event::KeyDown(keycode), + ElementState::Released => Event::KeyUp(keycode), + } + }) + } + WindowEvent::CloseRequested => Some(Event::Quit), + WindowEvent::Resized(new_size) => { + let logical_size = vec2i(new_size.width as i32, new_size.height as i32); + let backing_scale_factor = + window.get_current_monitor().get_hidpi_factor() as f32; + Some(Event::WindowResized(WindowSize { + logical_size, + backing_scale_factor, + })) + } + _ => None, + } + } + _ => None, + } +} \ No newline at end of file diff --git a/examples/canvas_glutin_minimal/src/main.rs b/examples/canvas_glutin_minimal/src/main.rs index d7e13961..747c57b2 100644 --- a/examples/canvas_glutin_minimal/src/main.rs +++ b/examples/canvas_glutin_minimal/src/main.rs @@ -51,7 +51,10 @@ fn main() { let mut renderer = Renderer::new(GLDevice::new(GLVersion::GL3, 0), &FilesystemResourceLoader::locate(), DestFramebuffer::full_window(window_size), - RendererOptions { background_color: Some(ColorF::white()) }); + RendererOptions { + background_color: Some(ColorF::white()), + ..RendererOptions::default() + }); // Make a canvas. We're going to draw a house. let font_context = CanvasFontContext::from_system_source(); diff --git a/examples/canvas_minimal/Cargo.toml b/examples/canvas_minimal/Cargo.toml index ca133cad..89bbe22c 100644 --- a/examples/canvas_minimal/Cargo.toml +++ b/examples/canvas_minimal/Cargo.toml @@ -6,8 +6,10 @@ edition = "2018" [dependencies] gl = "0.14" -sdl2 = "0.33" -sdl2-sys = "0.33" + +[dependencies.euclid] +version = "0.20" +features = [] [dependencies.pathfinder_canvas] path = "../../canvas" @@ -32,3 +34,10 @@ path = "../../renderer" [dependencies.pathfinder_resources] path = "../../resources" + +[dependencies.surfman] +version = "0.2" +features = ["sm-winit"] + +[dependencies.winit] +version = "<0.19.4" # 0.19.4 causes build errors https://github.com/rust-windowing/winit/pull/1105 diff --git a/examples/canvas_minimal/src/main.rs b/examples/canvas_minimal/src/main.rs index cfe26154..c80dbf98 100644 --- a/examples/canvas_minimal/src/main.rs +++ b/examples/canvas_minimal/src/main.rs @@ -8,6 +8,7 @@ // option. This file may not be copied, modified, or distributed // except according to those terms. +use euclid::default::Size2D; use pathfinder_canvas::{Canvas, CanvasFontContext, Path2D}; use pathfinder_color::ColorF; use pathfinder_geometry::rect::RectF; @@ -19,41 +20,69 @@ use pathfinder_renderer::gpu::options::{DestFramebuffer, RendererOptions}; use pathfinder_renderer::gpu::renderer::Renderer; use pathfinder_renderer::options::BuildOptions; use pathfinder_resources::embedded::EmbeddedResourceLoader; -use sdl2::event::Event; -use sdl2::keyboard::Keycode; -use sdl2::video::GLProfile; +use surfman::{Connection, ContextAttributeFlags, ContextAttributes, GLVersion as SurfmanGLVersion}; +use surfman::{SurfaceAccess, SurfaceType}; +use winit::dpi::LogicalSize; +use winit::{ControlFlow, Event, EventsLoop, WindowBuilder, WindowEvent}; fn main() { - // Set up SDL2. - let sdl_context = sdl2::init().unwrap(); - let video = sdl_context.video().unwrap(); - - // Make sure we have at least a GL 3.0 context. Pathfinder requires this. - let gl_attributes = video.gl_attr(); - gl_attributes.set_context_profile(GLProfile::Core); - gl_attributes.set_context_version(3, 3); - // Open a window. - let window_size = vec2i(640, 480); - let window = video.window("Minimal example", window_size.x() as u32, window_size.y() as u32) - .opengl() - .build() - .unwrap(); + let mut event_loop = EventsLoop::new(); + let window_size = Size2D::new(640, 480); + let logical_size = LogicalSize::new(window_size.width as f64, window_size.height as f64); + let window = WindowBuilder::new().with_title("Minimal example") + .with_dimensions(logical_size) + .build(&event_loop) + .unwrap(); + window.show(); - // Create the GL context, and make it current. - let gl_context = window.gl_create_context().unwrap(); - gl::load_with(|name| video.gl_get_proc_address(name) as *const _); - window.gl_make_current(&gl_context).unwrap(); + // Create a `surfman` device. On a multi-GPU system, we'll request the low-power integrated + // GPU. + let connection = Connection::from_winit_window(&window).unwrap(); + let native_widget = connection.create_native_widget_from_winit_window(&window).unwrap(); + let adapter = connection.create_low_power_adapter().unwrap(); + let mut device = connection.create_device(&adapter).unwrap(); + + // Request an OpenGL 3.x context. Pathfinder requires this. + let context_attributes = ContextAttributes { + version: SurfmanGLVersion::new(3, 0), + flags: ContextAttributeFlags::ALPHA, + }; + let context_descriptor = device.create_context_descriptor(&context_attributes).unwrap(); + + // Make the OpenGL context via `surfman`, and load OpenGL functions. + let surface_type = SurfaceType::Widget { native_widget }; + let mut context = device.create_context(&context_descriptor).unwrap(); + let surface = device.create_surface(&context, SurfaceAccess::GPUOnly, surface_type) + .unwrap(); + device.bind_surface_to_context(&mut context, surface).unwrap(); + device.make_context_current(&context).unwrap(); + gl::load_with(|symbol_name| device.get_proc_address(&context, symbol_name)); + + // Get the real size of the window, taking HiDPI into account. + let hidpi_factor = window.get_current_monitor().get_hidpi_factor(); + let physical_size = logical_size.to_physical(hidpi_factor); + let framebuffer_size = vec2i(physical_size.width as i32, physical_size.height as i32); + + // Create a Pathfinder GL device. + let default_framebuffer = device.context_surface_info(&context) + .unwrap() + .unwrap() + .framebuffer_object; + let pathfinder_device = GLDevice::new(GLVersion::GL3, default_framebuffer); // Create a Pathfinder renderer. - let mut renderer = Renderer::new(GLDevice::new(GLVersion::GL3, 0), + let mut renderer = Renderer::new(pathfinder_device, &EmbeddedResourceLoader::new(), - DestFramebuffer::full_window(window_size), - RendererOptions { background_color: Some(ColorF::white()) }); + DestFramebuffer::full_window(framebuffer_size), + RendererOptions { + background_color: Some(ColorF::white()), + ..RendererOptions::default() + }); // Make a canvas. We're going to draw a house. let font_context = CanvasFontContext::from_system_source(); - let mut canvas = Canvas::new(window_size.to_f32()).get_context_2d(font_context); + let mut canvas = Canvas::new(framebuffer_size.to_f32()).get_context_2d(font_context); // Set line width. canvas.set_line_width(10.0); @@ -75,14 +104,23 @@ fn main() { // Render the canvas to screen. let scene = SceneProxy::from_scene(canvas.into_canvas().into_scene(), RayonExecutor); scene.build_and_render(&mut renderer, BuildOptions::default()); - window.gl_swap_window(); + + // Present the surface. + let mut surface = device.unbind_surface_from_context(&mut context).unwrap().unwrap(); + device.present_surface(&mut context, &mut surface).unwrap(); + device.bind_surface_to_context(&mut context, surface).unwrap(); // Wait for a keypress. - let mut event_pump = sdl_context.event_pump().unwrap(); - loop { - match event_pump.wait_event() { - Event::Quit {..} | Event::KeyDown { keycode: Some(Keycode::Escape), .. } => return, - _ => {} + event_loop.run_forever(|event| { + match event { + Event::WindowEvent { event: WindowEvent::CloseRequested, .. } | + Event::WindowEvent { event: WindowEvent::KeyboardInput { .. }, .. } => { + ControlFlow::Break + } + _ => ControlFlow::Continue, } - } + }); + + // Clean up. + drop(device.destroy_context(&mut context)); } diff --git a/examples/canvas_moire/Cargo.toml b/examples/canvas_moire/Cargo.toml index 3f9a1f9a..7fc402c3 100644 --- a/examples/canvas_moire/Cargo.toml +++ b/examples/canvas_moire/Cargo.toml @@ -6,8 +6,10 @@ edition = "2018" [dependencies] gl = "0.14" -sdl2 = "0.33" -sdl2-sys = "0.33" + +[dependencies.euclid] +version = "0.20" +features = [] [dependencies.pathfinder_canvas] features = ["pf-text"] @@ -33,3 +35,11 @@ path = "../../renderer" [dependencies.pathfinder_resources] path = "../../resources" + +[dependencies.surfman] +# version = "0.2" +path = "/Users/pcwalton/Source/surfman/surfman" +features = ["sm-winit"] + +[dependencies.winit] +version = "<0.19.4" # 0.19.4 causes build errors https://github.com/rust-windowing/winit/pull/1105 diff --git a/examples/canvas_moire/src/main.rs b/examples/canvas_moire/src/main.rs index 888c5f87..dce7b317 100644 --- a/examples/canvas_moire/src/main.rs +++ b/examples/canvas_moire/src/main.rs @@ -8,6 +8,7 @@ // option. This file may not be copied, modified, or distributed // except according to those terms. +use euclid::default::Size2D; use pathfinder_canvas::{Canvas, CanvasFontContext, CanvasRenderingContext2D, FillStyle, Path2D}; use pathfinder_color::{ColorF, ColorU}; use pathfinder_geometry::vector::{Vector2F, Vector2I, vec2f, vec2i}; @@ -18,11 +19,12 @@ use pathfinder_renderer::gpu::options::{DestFramebuffer, RendererOptions}; use pathfinder_renderer::gpu::renderer::Renderer; use pathfinder_renderer::options::BuildOptions; use pathfinder_resources::fs::FilesystemResourceLoader; -use sdl2::event::Event; -use sdl2::keyboard::Keycode; -use sdl2::video::GLProfile; use std::f32::consts::PI; use std::f32; +use surfman::{Connection, ContextAttributeFlags, ContextAttributes, GLVersion as SurfmanGLVersion}; +use surfman::{SurfaceAccess, SurfaceType}; +use winit::dpi::LogicalSize; +use winit::{Event, EventsLoop, WindowBuilder, WindowEvent}; const VELOCITY: f32 = 0.02; const OUTER_RADIUS: f32 = 64.0; @@ -37,50 +39,79 @@ const CIRCLE_THICKNESS: f32 = 16.0; const COLOR_CYCLE_SPEED: f32 = 0.0025; fn main() { - // Set up SDL2. - let sdl_context = sdl2::init().unwrap(); - let video = sdl_context.video().unwrap(); - - // Make sure we have at least a GL 3.0 context. Pathfinder requires this. - let gl_attributes = video.gl_attr(); - gl_attributes.set_context_profile(GLProfile::Core); - gl_attributes.set_context_version(3, 3); - // Open a window. - let window_size = vec2i(1067, 800); - let window = video.window("Moire example", window_size.x() as u32, window_size.y() as u32) - .opengl() - .allow_highdpi() - .build() - .unwrap(); - let mut event_pump = sdl_context.event_pump().unwrap(); + let mut event_loop = EventsLoop::new(); + let window_size = Size2D::new(1067, 800); + let logical_size = LogicalSize::new(window_size.width as f64, window_size.height as f64); + let window = WindowBuilder::new().with_title("Moire example") + .with_dimensions(logical_size) + .build(&event_loop) + .unwrap(); + window.show(); - // Get the real window size (for HiDPI). - let (drawable_width, drawable_height) = window.drawable_size(); - let drawable_size = vec2i(drawable_width as i32, drawable_height as i32); + // Create a `surfman` device. On a multi-GPU system, we'll request the low-power integrated + // GPU. + let connection = Connection::from_winit_window(&window).unwrap(); + let native_widget = connection.create_native_widget_from_winit_window(&window).unwrap(); + let adapter = connection.create_low_power_adapter().unwrap(); + let mut device = connection.create_device(&adapter).unwrap(); - // Create the GL context, and make it current. - let gl_context = window.gl_create_context().unwrap(); - gl::load_with(|name| video.gl_get_proc_address(name) as *const _); - window.gl_make_current(&gl_context).unwrap(); + // Request an OpenGL 3.x context. Pathfinder requires this. + let context_attributes = ContextAttributes { + version: SurfmanGLVersion::new(3, 0), + flags: ContextAttributeFlags::ALPHA, + }; + let context_descriptor = device.create_context_descriptor(&context_attributes).unwrap(); + + // Make the OpenGL context via `surfman`, and load OpenGL functions. + let surface_type = SurfaceType::Widget { native_widget }; + let mut gl_context = device.create_context(&context_descriptor).unwrap(); + let surface = device.create_surface(&gl_context, SurfaceAccess::GPUOnly, surface_type) + .unwrap(); + device.bind_surface_to_context(&mut gl_context, surface).unwrap(); + device.make_context_current(&gl_context).unwrap(); + gl::load_with(|symbol_name| device.get_proc_address(&gl_context, symbol_name)); + + // Get the real size of the window, taking HiDPI into account. + let hidpi_factor = window.get_current_monitor().get_hidpi_factor(); + let physical_size = logical_size.to_physical(hidpi_factor); + let framebuffer_size = vec2i(physical_size.width as i32, physical_size.height as i32); + + // Create a Pathfinder GL device. + let default_framebuffer = device.context_surface_info(&gl_context) + .unwrap() + .unwrap() + .framebuffer_object; + let pathfinder_device = GLDevice::new(GLVersion::GL3, default_framebuffer); // Create our renderers. - let renderer = Renderer::new(GLDevice::new(GLVersion::GL3, 0), + let renderer = Renderer::new(pathfinder_device, &FilesystemResourceLoader::locate(), - DestFramebuffer::full_window(drawable_size), - RendererOptions { background_color: Some(ColorF::white()) }); - let mut moire_renderer = MoireRenderer::new(renderer, window_size, drawable_size); + DestFramebuffer::full_window(framebuffer_size), + RendererOptions { + background_color: Some(ColorF::white()), + ..RendererOptions::default() + }); + let window_size = vec2i(window_size.width, window_size.height); + let mut moire_renderer = MoireRenderer::new(renderer, window_size, framebuffer_size); // Enter main render loop. - loop { + let mut exit = false; + while !exit { moire_renderer.render(); - window.gl_swap_window(); - match event_pump.poll_event() { - Some(Event::Quit {..}) | - Some(Event::KeyDown { keycode: Some(Keycode::Escape), .. }) => return, - _ => {} - } + // Present the rendered canvas via `surfman`. + let mut surface = device.unbind_surface_from_context(&mut gl_context).unwrap().unwrap(); + device.present_surface(&mut gl_context, &mut surface).unwrap(); + device.bind_surface_to_context(&mut gl_context, surface).unwrap(); + + event_loop.poll_events(|event| { + match event { + Event::WindowEvent { event: WindowEvent::CloseRequested, .. } | + Event::WindowEvent { event: WindowEvent::KeyboardInput { .. }, .. } => exit = true, + _ => {} + } + }); } } @@ -124,7 +155,10 @@ impl MoireRenderer { let inner_center = window_center + vec2f(1.0, sin_time) * (cos_time * INNER_RADIUS); // Clear to background color. - self.renderer.set_options(RendererOptions { background_color: Some(background_color) }); + self.renderer.set_options(RendererOptions { + background_color: Some(background_color), + ..RendererOptions::default() + }); // Make a canvas. let mut canvas = diff --git a/examples/canvas_nanovg/Cargo.toml b/examples/canvas_nanovg/Cargo.toml index fdafb3fc..837a578d 100644 --- a/examples/canvas_nanovg/Cargo.toml +++ b/examples/canvas_nanovg/Cargo.toml @@ -8,8 +8,10 @@ edition = "2018" arrayvec = "0.5" font-kit = "0.6" gl = "0.14" -sdl2 = "0.33" -sdl2-sys = "0.33" + +[dependencies.euclid] +version = "0.20" +features = [] [dependencies.image] version = "0.23" @@ -48,5 +50,13 @@ path = "../../resources" [dependencies.pathfinder_simd] path = "../../simd" +[dependencies.surfman] +git = "https://github.com/pcwalton/surfman" +rev = "bc29e7ae88ca7dd64d3b1c7e185604693290207f" +features = ["sm-winit"] + +[dependencies.winit] +version = "<0.19.4" # 0.19.4 causes build errors https://github.com/rust-windowing/winit/pull/1105 + [target.'cfg(not(windows))'.dependencies] jemallocator = "0.3" diff --git a/examples/canvas_nanovg/src/main.rs b/examples/canvas_nanovg/src/main.rs index f7c49a20..0a238163 100644 --- a/examples/canvas_nanovg/src/main.rs +++ b/examples/canvas_nanovg/src/main.rs @@ -9,6 +9,7 @@ // except according to those terms. use arrayvec::ArrayVec; +use euclid::default::Size2D; use font_kit::handle::Handle; use font_kit::sources::mem::MemSource; use image; @@ -35,14 +36,15 @@ use pathfinder_renderer::options::BuildOptions; use pathfinder_resources::ResourceLoader; use pathfinder_resources::fs::FilesystemResourceLoader; use pathfinder_simd::default::F32x2; -use sdl2::event::Event; -use sdl2::keyboard::Keycode; -use sdl2::video::GLProfile; use std::collections::VecDeque; use std::f32::consts::PI; use std::iter; use std::sync::Arc; use std::time::Instant; +use surfman::{Connection, ContextAttributeFlags, ContextAttributes, GLVersion as SurfmanGLVersion}; +use surfman::{SurfaceAccess, SurfaceType}; +use winit::dpi::LogicalSize; +use winit::{Event, EventsLoop, WindowBuilder, WindowEvent}; #[cfg(not(windows))] use jemallocator; @@ -1460,33 +1462,43 @@ impl DemoData { } fn main() { - // Set up SDL2. - let sdl_context = sdl2::init().unwrap(); - let video = sdl_context.video().unwrap(); - - // Make sure we have at least a GL 3.0 context. Pathfinder requires this. - let gl_attributes = video.gl_attr(); - gl_attributes.set_context_profile(GLProfile::Core); - gl_attributes.set_context_version(3, 3); - // Open a window. - let window_size = vec2i(WINDOW_WIDTH, WINDOW_HEIGHT); - let window = - video.window("NanoVG example port", window_size.x() as u32, window_size.y() as u32) - .opengl() - .allow_highdpi() - .build() - .unwrap(); + let mut event_loop = EventsLoop::new(); + let window_size = Size2D::new(WINDOW_WIDTH, WINDOW_HEIGHT); + let logical_size = LogicalSize::new(window_size.width as f64, window_size.height as f64); + let window = WindowBuilder::new().with_title("NanoVG example port") + .with_dimensions(logical_size) + .build(&event_loop) + .unwrap(); + window.show(); - // Create the GL context, and make it current. - let gl_context = window.gl_create_context().unwrap(); - gl::load_with(|name| video.gl_get_proc_address(name) as *const _); - window.gl_make_current(&gl_context).unwrap(); + // Create a `surfman` device. On a multi-GPU system, we'll request the low-power integrated + // GPU. + let connection = Connection::from_winit_window(&window).unwrap(); + let native_widget = connection.create_native_widget_from_winit_window(&window).unwrap(); + let adapter = connection.create_low_power_adapter().unwrap(); + let mut device = connection.create_device(&adapter).unwrap(); - // Get the real window size (for HiDPI). - let (drawable_width, drawable_height) = window.drawable_size(); - let drawable_size = vec2i(drawable_width as i32, drawable_height as i32); - let hidpi_factor = drawable_size.x() as f32 / window_size.x() as f32; + // Request an OpenGL 3.x context. Pathfinder requires this. + let context_attributes = ContextAttributes { + version: SurfmanGLVersion::new(3, 0), + flags: ContextAttributeFlags::ALPHA, + }; + let context_descriptor = device.create_context_descriptor(&context_attributes).unwrap(); + + // Make the OpenGL context via `surfman`, and load OpenGL functions. + let surface_type = SurfaceType::Widget { native_widget }; + let mut gl_context = device.create_context(&context_descriptor).unwrap(); + let surface = device.create_surface(&gl_context, SurfaceAccess::GPUOnly, surface_type) + .unwrap(); + device.bind_surface_to_context(&mut gl_context, surface).unwrap(); + device.make_context_current(&gl_context).unwrap(); + gl::load_with(|symbol_name| device.get_proc_address(&gl_context, symbol_name)); + + // Get the real size of the window, taking HiDPI into account. + let hidpi_factor = window.get_current_monitor().get_hidpi_factor(); + let physical_size = logical_size.to_physical(hidpi_factor); + let framebuffer_size = vec2i(physical_size.width as i32, physical_size.height as i32); // Load demo data. let resources = FilesystemResourceLoader::locate(); @@ -1497,12 +1509,20 @@ fn main() { ]; let demo_data = DemoData::load(&resources); + // Create a Pathfinder GL device. + let default_framebuffer = device.context_surface_info(&gl_context) + .unwrap() + .unwrap() + .framebuffer_object; + let pathfinder_device = GLDevice::new(GLVersion::GL3, default_framebuffer); + // Create a Pathfinder renderer. - let mut renderer = Renderer::new(GLDevice::new(GLVersion::GL3, 0), + let mut renderer = Renderer::new(pathfinder_device, &resources, - DestFramebuffer::full_window(drawable_size), + DestFramebuffer::full_window(framebuffer_size), RendererOptions { background_color: Some(rgbf(0.3, 0.3, 0.32)), + ..RendererOptions::default() }); // Initialize font state. @@ -1510,7 +1530,6 @@ fn main() { let font_context = CanvasFontContext::new(font_source.clone()); // Initialize general state. - let mut event_pump = sdl_context.event_pump().unwrap(); let mut mouse_position = Vector2F::zero(); let start_time = Instant::now(); @@ -1520,21 +1539,23 @@ fn main() { let mut gpu_graph = PerfGraph::new(GraphStyle::MS, "GPU Time"); // Enter the main loop. - loop { + let mut exit = false; + while !exit { // Make a canvas. - let mut context = Canvas::new(drawable_size.to_f32()).get_context_2d(font_context.clone()); + let mut context = + Canvas::new(framebuffer_size.to_f32()).get_context_2d(font_context.clone()); // Start performance timing. let frame_start_time = Instant::now(); let frame_start_elapsed_time = (frame_start_time - start_time).as_secs_f32(); // Render the demo. - context.scale(hidpi_factor); + context.scale(hidpi_factor as f32); render_demo(&mut context, mouse_position, - window_size.to_f32(), + vec2f(WINDOW_WIDTH as f32, WINDOW_HEIGHT as f32), frame_start_elapsed_time, - hidpi_factor, + hidpi_factor as f32, &demo_data); // Render performance graphs. @@ -1547,7 +1568,11 @@ fn main() { let canvas = context.into_canvas(); let scene = SceneProxy::from_scene(canvas.into_scene(), RayonExecutor); scene.build_and_render(&mut renderer, BuildOptions::default()); - window.gl_swap_window(); + + // Present the rendered canvas via `surfman`. + let mut surface = device.unbind_surface_from_context(&mut gl_context).unwrap().unwrap(); + device.present_surface(&mut gl_context, &mut surface).unwrap(); + device.bind_surface_to_context(&mut gl_context, surface).unwrap(); // Add stats to performance graphs. if let Some(gpu_time) = renderer.shift_rendering_time() { @@ -1558,12 +1583,18 @@ fn main() { gpu_graph.push(gpu_time); } - for event in event_pump.poll_iter() { + event_loop.poll_events(|event| { match event { - Event::Quit {..} | Event::KeyDown { keycode: Some(Keycode::Escape), .. } => return, - Event::MouseMotion { x, y, .. } => mouse_position = vec2i(x, y).to_f32(), + Event::WindowEvent { event: WindowEvent::CloseRequested, .. } | + Event::WindowEvent { event: WindowEvent::KeyboardInput { .. }, .. } => exit = true, + Event::WindowEvent { event: WindowEvent::CursorMoved { position, .. }, .. } => { + mouse_position = vec2f(position.x as f32, position.y as f32); + } _ => {} } - } + }); } + + // Clean up. + drop(device.destroy_context(&mut gl_context)); } diff --git a/examples/canvas_text/src/main.rs b/examples/canvas_text/src/main.rs index 3f1a4ba1..85603a35 100644 --- a/examples/canvas_text/src/main.rs +++ b/examples/canvas_text/src/main.rs @@ -53,7 +53,10 @@ fn main() { let mut renderer = Renderer::new(GLDevice::new(GLVersion::GL3, 0), &resource_loader, DestFramebuffer::full_window(window_size), - RendererOptions { background_color: Some(ColorF::white()) }); + RendererOptions { + background_color: Some(ColorF::white()), + ..RendererOptions::default() + }); // Load a font. let font_data = Arc::new(resource_loader.slurp("fonts/Overpass-Regular.otf").unwrap()); diff --git a/examples/swf_basic/src/main.rs b/examples/swf_basic/src/main.rs index 11c15424..239ddb53 100644 --- a/examples/swf_basic/src/main.rs +++ b/examples/swf_basic/src/main.rs @@ -102,7 +102,10 @@ fn main() { GLDevice::new(GLVersion::GL3, 0), &resource_loader, DestFramebuffer::full_window(pixel_size), - RendererOptions { background_color: Some(stage.background_color()) } + RendererOptions { + background_color: Some(stage.background_color()), + ..RendererOptions::default() + } ); // Clear to swf stage background color. let mut scene = Scene::new(); diff --git a/gl/src/lib.rs b/gl/src/lib.rs index 4c6883ed..1c3173ac 100644 --- a/gl/src/lib.rs +++ b/gl/src/lib.rs @@ -13,13 +13,13 @@ #[macro_use] extern crate log; -use gl::types::{GLboolean, GLchar, GLenum, GLfloat, GLint, GLsizei, GLsizeiptr, GLsync}; +use gl::types::{GLboolean, GLchar, GLenum, GLfloat, GLint, GLintptr, GLsizei, GLsizeiptr, GLsync}; use gl::types::{GLuint, GLvoid}; use half::f16; use pathfinder_geometry::rect::RectI; use pathfinder_geometry::vector::Vector2I; -use pathfinder_gpu::{BlendFactor, BlendOp, BufferData, BufferTarget, BufferUploadMode, ClearOps}; -use pathfinder_gpu::{DepthFunc, Device, Primitive, RenderOptions, RenderState, RenderTarget}; +use pathfinder_gpu::{BlendFactor, BlendOp, BufferData, BufferTarget, BufferUploadMode, ClearOps, ComputeDimensions, ComputeState}; +use pathfinder_gpu::{DepthFunc, Device, ImageAccess, ImageBinding, Primitive, ProgramKind, RenderOptions, RenderState, RenderTarget}; use pathfinder_gpu::{ShaderKind, StencilFunc, TextureData, TextureDataRef, TextureFormat}; use pathfinder_gpu::{TextureSamplingFlags, UniformData, VertexAttrClass}; use pathfinder_gpu::{VertexAttrDescriptor, VertexAttrType}; @@ -71,6 +71,18 @@ impl GLDevice { self.set_render_options(&render_state.options); } + fn set_compute_state(&self, compute_state: &ComputeState) { + self.use_program(compute_state.program); + for (texture_unit, texture) in compute_state.textures.iter().enumerate() { + self.bind_texture(texture, texture_unit as u32); + } + for (image_unit, image) in compute_state.images.iter().enumerate() { + self.bind_image(image, image_unit as u32); + } + + compute_state.uniforms.iter().for_each(|(uniform, data)| self.set_uniform(uniform, data)); + } + fn set_render_options(&self, render_options: &RenderOptions) { unsafe { // Set blend. @@ -165,7 +177,7 @@ impl GLDevice { UniformData::Vec4(data) => { gl::Uniform4f(uniform.location, data.x(), data.y(), data.z(), data.w()); ck(); } - UniformData::TextureUnit(unit) => { + UniformData::TextureUnit(unit) | UniformData::ImageUnit(unit) => { gl::Uniform1i(uniform.location, unit as GLint); ck(); } } @@ -194,6 +206,19 @@ impl GLDevice { self.unbind_vertex_array(); } + fn reset_compute_state(&self, compute_state: &ComputeState) { + for image_unit in 0..(compute_state.images.len() as u32) { + self.unbind_image(image_unit); + } + for texture_unit in 0..(compute_state.textures.len() as u32) { + self.unbind_texture(texture_unit); + } + for (uniform, data) in compute_state.uniforms { + self.unset_uniform(uniform, data); + } + self.unuse_program(); + } + fn reset_render_options(&self, render_options: &RenderOptions) { unsafe { if render_options.blend.is_some() { @@ -216,9 +241,11 @@ impl GLDevice { impl Device for GLDevice { type Buffer = GLBuffer; + type Fence = GLFence; type Framebuffer = GLFramebuffer; type Program = GLProgram; type Shader = GLShader; + type StorageBuffer = GLStorageBuffer; type Texture = GLTexture; type TextureDataReceiver = GLTextureDataReceiver; type TimerQuery = GLTimerQuery; @@ -277,8 +304,9 @@ impl Device for GLDevice { let source = output; let gl_shader_kind = match kind { - ShaderKind::Vertex => gl::VERTEX_SHADER, + ShaderKind::Vertex => gl::VERTEX_SHADER, ShaderKind::Fragment => gl::FRAGMENT_SHADER, + ShaderKind::Compute => gl::COMPUTE_SHADER, }; unsafe { @@ -310,14 +338,23 @@ impl Device for GLDevice { fn create_program_from_shaders(&self, _resources: &dyn ResourceLoader, name: &str, - vertex_shader: GLShader, - fragment_shader: GLShader) + shaders: ProgramKind) -> GLProgram { let gl_program; unsafe { gl_program = gl::CreateProgram(); ck(); - gl::AttachShader(gl_program, vertex_shader.gl_shader); ck(); - gl::AttachShader(gl_program, fragment_shader.gl_shader); ck(); + match shaders { + ProgramKind::Raster { + vertex: ref vertex_shader, + fragment: ref fragment_shader, + } => { + gl::AttachShader(gl_program, vertex_shader.gl_shader); ck(); + gl::AttachShader(gl_program, fragment_shader.gl_shader); ck(); + } + ProgramKind::Compute(ref compute_shader) => { + gl::AttachShader(gl_program, compute_shader.gl_shader); ck(); + } + } gl::LinkProgram(gl_program); ck(); let mut link_status = 0; @@ -335,7 +372,17 @@ impl Device for GLDevice { } } - GLProgram { gl_program, vertex_shader, fragment_shader } + match shaders { + ProgramKind::Raster { vertex: vertex_shader, fragment: fragment_shader } => { + GLProgram { gl_program, vertex_shader, fragment_shader } + } + ProgramKind::Compute(_) => unimplemented!(), + } + } + + #[inline] + fn set_compute_program_local_size(&self, _: &mut Self::Program, _: ComputeDimensions) { + // This does nothing on OpenGL, since the local size is set in the shader. } #[inline] @@ -367,6 +414,10 @@ impl Device for GLDevice { GLUniform { location } } + fn get_storage_buffer(&self, _: &Self::Program, _: &str, binding: u32) -> GLStorageBuffer { + GLStorageBuffer { location: binding as GLint } + } + fn configure_vertex_attr(&self, vertex_array: &GLVertexArray, attr: &GLVertexAttr, @@ -424,35 +475,47 @@ impl Device for GLDevice { GLFramebuffer { gl_framebuffer, texture } } - fn create_buffer(&self) -> GLBuffer { + fn create_buffer(&self, mode: BufferUploadMode) -> GLBuffer { unsafe { let mut gl_buffer = 0; gl::GenBuffers(1, &mut gl_buffer); ck(); - GLBuffer { gl_buffer } + GLBuffer { gl_buffer, mode } } } fn allocate_buffer(&self, buffer: &GLBuffer, data: BufferData, - target: BufferTarget, - mode: BufferUploadMode) { - let target = match target { - BufferTarget::Vertex => gl::ARRAY_BUFFER, - BufferTarget::Index => gl::ELEMENT_ARRAY_BUFFER, - }; + target: BufferTarget) { + let target = target.to_gl_target(); let (ptr, len) = match data { BufferData::Uninitialized(len) => (ptr::null(), len), BufferData::Memory(buffer) => (buffer.as_ptr() as *const GLvoid, buffer.len()), }; let len = (len * mem::size_of::()) as GLsizeiptr; - let usage = mode.to_gl_usage(); + let usage = buffer.mode.to_gl_usage(); unsafe { gl::BindBuffer(target, buffer.gl_buffer); ck(); gl::BufferData(target, len, ptr, usage); ck(); } } + fn upload_to_buffer(&self, + buffer: &Self::Buffer, + position: usize, + data: &[T], + target: BufferTarget) { + let target = target.to_gl_target(); + let len = (data.len() * mem::size_of::()) as GLsizeiptr; + unsafe { + gl::BindBuffer(target, buffer.gl_buffer); ck(); + gl::BufferSubData(target, + position as GLintptr, + len, + data.as_ptr() as *const GLvoid); ck(); + } + } + #[inline] fn framebuffer_texture<'f>(&self, framebuffer: &'f Self::Framebuffer) -> &'f Self::Texture { &framebuffer.texture @@ -627,6 +690,14 @@ impl Device for GLDevice { self.reset_render_state(render_state); } + fn dispatch_compute(&self, dimensions: ComputeDimensions, compute_state: &ComputeState) { + self.set_compute_state(compute_state); + unsafe { + gl::DispatchCompute(dimensions.x, dimensions.y, dimensions.z); ck(); + } + self.reset_compute_state(compute_state); + } + #[inline] fn create_timer_query(&self) -> GLTimerQuery { let mut query = GLTimerQuery { gl_query: 0 }; @@ -712,10 +783,24 @@ impl Device for GLDevice { let suffix = match kind { ShaderKind::Vertex => 'v', ShaderKind::Fragment => 'f', + ShaderKind::Compute => 'c', }; let path = format!("shaders/gl3/{}.{}s.glsl", name, suffix); self.create_shader_from_source(name, &resources.slurp(&path).unwrap(), kind) } + + fn add_fence(&self) -> Self::Fence { + unsafe { + let gl_sync = gl::FenceSync(gl::SYNC_GPU_COMMANDS_COMPLETE, 0); ck(); + GLFence { gl_sync } + } + } + + fn wait_for_fence(&self, fence: &Self::Fence) { + unsafe { + gl::ClientWaitSync(fence.gl_sync, gl::SYNC_FLUSH_COMMANDS_BIT, 0); ck(); + } + } } impl GLDevice { @@ -752,6 +837,24 @@ impl GLDevice { } } + fn bind_image(&self, binding: &ImageBinding, unit: u32) { + unsafe { + gl::BindImageTexture(unit, + binding.texture.gl_texture, + 0, + gl::FALSE, + 0, + binding.access.to_gl_access(), + binding.texture.format.gl_internal_format() as GLenum); ck(); + } + } + + fn unbind_image(&self, unit: u32) { + unsafe { + gl::BindImageTexture(unit, 0, 0, gl::FALSE, 0, 0, 0); ck(); + } + } + fn use_program(&self, program: &GLProgram) { unsafe { gl::UseProgram(program.gl_program); ck(); @@ -933,6 +1036,18 @@ impl GLVertexAttr { } } +pub struct GLFence { + pub gl_sync: GLsync, +} + +impl Drop for GLFence { + fn drop(&mut self) { + unsafe { + gl::DeleteSync(self.gl_sync); ck(); + } + } +} + pub struct GLFramebuffer { pub gl_framebuffer: GLuint, pub texture: GLTexture, @@ -948,6 +1063,7 @@ impl Drop for GLFramebuffer { pub struct GLBuffer { pub gl_buffer: GLuint, + pub mode: BufferUploadMode, } impl Drop for GLBuffer { @@ -963,6 +1079,11 @@ pub struct GLUniform { location: GLint, } +#[derive(Debug)] +pub struct GLStorageBuffer { + location: GLint, +} + pub struct GLProgram { pub gl_program: GLuint, #[allow(dead_code)] @@ -1063,6 +1184,7 @@ impl BufferTargetExt for BufferTarget { match self { BufferTarget::Vertex => gl::ARRAY_BUFFER, BufferTarget::Index => gl::ELEMENT_ARRAY_BUFFER, + BufferTarget::Storage => gl::SHADER_STORAGE_BUFFER, } } } @@ -1093,6 +1215,20 @@ impl DepthFuncExt for DepthFunc { } } +trait ImageAccessExt { + fn to_gl_access(self) -> GLenum; +} + +impl ImageAccessExt for ImageAccess { + fn to_gl_access(self) -> GLenum { + match self { + ImageAccess::Read => gl::READ_ONLY, + ImageAccess::Write => gl::WRITE_ONLY, + ImageAccess::ReadWrite => gl::READ_WRITE, + } + } +} + trait PrimitiveExt { fn to_gl_primitive(self) -> GLuint; } @@ -1192,6 +1328,8 @@ pub enum GLVersion { GL3 = 0, /// OpenGL ES 3.0+. GLES3 = 1, + /// OpenGL 4.3+, core profile. + GL4_3 = 2, } impl GLVersion { @@ -1199,6 +1337,7 @@ impl GLVersion { match *self { GLVersion::GL3 => "330", GLVersion::GLES3 => "300 es", + GLVersion::GL4_3 => "430", } } } diff --git a/gpu/src/lib.rs b/gpu/src/lib.rs index 5961c6cf..47009731 100644 --- a/gpu/src/lib.rs +++ b/gpu/src/lib.rs @@ -26,9 +26,11 @@ use std::time::Duration; pub trait Device: Sized { type Buffer; + type Fence; type Framebuffer; type Program; type Shader; + type StorageBuffer; type Texture; type TextureDataReceiver; type TimerQuery; @@ -44,15 +46,18 @@ pub trait Device: Sized { fn create_shader_from_source(&self, name: &str, source: &[u8], kind: ShaderKind) -> Self::Shader; fn create_vertex_array(&self) -> Self::VertexArray; - fn create_program_from_shaders( - &self, - resources: &dyn ResourceLoader, - name: &str, - vertex_shader: Self::Shader, - fragment_shader: Self::Shader, - ) -> Self::Program; + fn create_program_from_shaders(&self, + resources: &dyn ResourceLoader, + name: &str, + shaders: ProgramKind) + -> Self::Program; + fn set_compute_program_local_size(&self, + program: &mut Self::Program, + local_size: ComputeDimensions); fn get_vertex_attr(&self, program: &Self::Program, name: &str) -> Option; fn get_uniform(&self, program: &Self::Program, name: &str) -> Self::Uniform; + fn get_storage_buffer(&self, program: &Self::Program, name: &str, binding: u32) + -> Self::StorageBuffer; fn bind_buffer(&self, vertex_array: &Self::VertexArray, buffer: &Self::Buffer, @@ -62,14 +67,16 @@ pub trait Device: Sized { attr: &Self::VertexAttr, descriptor: &VertexAttrDescriptor); fn create_framebuffer(&self, texture: Self::Texture) -> Self::Framebuffer; - fn create_buffer(&self) -> Self::Buffer; - fn allocate_buffer( - &self, - buffer: &Self::Buffer, - data: BufferData, - target: BufferTarget, - mode: BufferUploadMode, - ); + fn create_buffer(&self, mode: BufferUploadMode) -> Self::Buffer; + fn allocate_buffer(&self, + buffer: &Self::Buffer, + data: BufferData, + target: BufferTarget); + fn upload_to_buffer(&self, + buffer: &Self::Buffer, + position: usize, + data: &[T], + target: BufferTarget); fn framebuffer_texture<'f>(&self, framebuffer: &'f Self::Framebuffer) -> &'f Self::Texture; fn destroy_framebuffer(&self, framebuffer: Self::Framebuffer) -> Self::Texture; fn texture_format(&self, texture: &Self::Texture) -> TextureFormat; @@ -86,6 +93,9 @@ pub trait Device: Sized { index_count: u32, instance_count: u32, render_state: &RenderState); + fn dispatch_compute(&self, dimensions: ComputeDimensions, state: &ComputeState); + fn add_fence(&self) -> Self::Fence; + fn wait_for_fence(&self, fence: &Self::Fence); fn create_timer_query(&self) -> Self::TimerQuery; fn begin_timer_query(&self, query: &Self::TimerQuery); fn end_timer_query(&self, query: &Self::TimerQuery); @@ -107,17 +117,30 @@ pub trait Device: Sized { &self, resources: &dyn ResourceLoader, program_name: &str, - vertex_shader_name: &str, - fragment_shader_name: &str, + shader_names: ProgramKind<&str>, ) -> Self::Program { - let vertex_shader = self.create_shader(resources, vertex_shader_name, ShaderKind::Vertex); - let fragment_shader = - self.create_shader(resources, fragment_shader_name, ShaderKind::Fragment); - self.create_program_from_shaders(resources, program_name, vertex_shader, fragment_shader) + let shaders = match shader_names { + ProgramKind::Raster { vertex, fragment } => { + ProgramKind::Raster { + vertex: self.create_shader(resources, vertex, ShaderKind::Vertex), + fragment: self.create_shader(resources, fragment, ShaderKind::Fragment), + } + } + ProgramKind::Compute(compute) => { + ProgramKind::Compute(self.create_shader(resources, compute, ShaderKind::Compute)) + } + }; + self.create_program_from_shaders(resources, program_name, shaders) } - fn create_program(&self, resources: &dyn ResourceLoader, name: &str) -> Self::Program { - self.create_program_from_shader_names(resources, name, name, name) + fn create_raster_program(&self, resources: &dyn ResourceLoader, name: &str) -> Self::Program { + let shaders = ProgramKind::Raster { vertex: name, fragment: name }; + self.create_program_from_shader_names(resources, name, shaders) + } + + fn create_compute_program(&self, resources: &dyn ResourceLoader, name: &str) -> Self::Program { + let shaders = ProgramKind::Compute(name); + self.create_program_from_shader_names(resources, name, shaders) } } @@ -149,6 +172,7 @@ pub enum BufferData<'a, T> { pub enum BufferTarget { Vertex, Index, + Storage, } #[derive(Clone, Copy, Debug)] @@ -161,6 +185,23 @@ pub enum BufferUploadMode { pub enum ShaderKind { Vertex, Fragment, + Compute, +} + +#[derive(Clone, Copy, Debug)] +pub enum ProgramKind { + Raster { + vertex: T, + fragment: T, + }, + Compute(T), +} + +#[derive(Clone, Copy, Debug, PartialEq)] +pub struct ComputeDimensions { + pub x: u32, + pub y: u32, + pub z: u32, } #[derive(Clone, Copy)] @@ -175,6 +216,7 @@ pub enum UniformData { Vec3([f32; 3]), Vec4(F32x4), TextureUnit(u32), + ImageUnit(u32), } #[derive(Clone, Copy)] @@ -191,10 +233,26 @@ pub struct RenderState<'a, D> where D: Device { pub primitive: Primitive, pub uniforms: &'a [(&'a D::Uniform, UniformData)], pub textures: &'a [&'a D::Texture], + pub images: &'a [ImageBinding<'a, D>], pub viewport: RectI, pub options: RenderOptions, } +#[derive(Clone)] +pub struct ComputeState<'a, D> where D: Device { + pub program: &'a D::Program, + pub uniforms: &'a [(&'a D::Uniform, UniformData)], + pub textures: &'a [&'a D::Texture], + pub images: &'a [ImageBinding<'a, D>], + pub storage_buffers: &'a [(&'a D::StorageBuffer, &'a D::Buffer)], +} + +#[derive(Clone, Debug)] +pub struct ImageBinding<'a, D> where D: Device { + pub texture: &'a D::Texture, + pub access: ImageAccess, +} + #[derive(Clone, Debug)] pub struct RenderOptions { pub blend: Option, @@ -408,6 +466,13 @@ bitflags! { } } +#[derive(Clone, Copy, Debug, PartialEq)] +pub enum ImageAccess { + Read, + Write, + ReadWrite, +} + impl<'a> TextureDataRef<'a> { #[doc(hidden)] pub fn check_and_extract_data_ptr(self, minimum_size: Vector2I, format: TextureFormat) diff --git a/metal/Cargo.toml b/metal/Cargo.toml index c3bec1cd..0bd817ee 100644 --- a/metal/Cargo.toml +++ b/metal/Cargo.toml @@ -13,9 +13,11 @@ bitflags = "1.0" byteorder = "1.3" block = "0.1" cocoa = "0.19" -core-foundation = "0.7" +core-foundation = "0.6" foreign-types = "0.3" half = "1.5" +io-surface = "0.12" +libc = "0.2" metal = "0.17" objc = "0.2" diff --git a/metal/src/lib.rs b/metal/src/lib.rs index 2075ada8..d2908e7c 100644 --- a/metal/src/lib.rs +++ b/metal/src/lib.rs @@ -24,45 +24,49 @@ use core_foundation::base::TCFType; use core_foundation::string::{CFString, CFStringRef}; use foreign_types::{ForeignType, ForeignTypeRef}; use half::f16; +use io_surface::IOSurfaceRef; +use libc::size_t; use metal::{self, Argument, ArgumentEncoder, Buffer, CommandBuffer, CommandBufferRef}; -use metal::{CommandQueue, CompileOptions, CoreAnimationDrawable, CoreAnimationDrawableRef}; +use metal::{CommandQueue, CompileOptions, ComputeCommandEncoderRef, ComputePipelineDescriptor}; +use metal::{ComputePipelineState, CoreAnimationDrawable, CoreAnimationDrawableRef}; use metal::{CoreAnimationLayer, CoreAnimationLayerRef, DepthStencilDescriptor, Function, Library}; -use metal::{MTLArgument, MTLArgumentEncoder, MTLBlendFactor, MTLBlendOperation, MTLClearColor}; -use metal::{MTLColorWriteMask, MTLCompareFunction, MTLDataType, MTLDevice, MTLFunctionType}; -use metal::{MTLIndexType, MTLLoadAction, MTLOrigin, MTLPixelFormat, MTLPrimitiveType, MTLRegion}; -use metal::{MTLRenderPipelineReflection, MTLRenderPipelineState, MTLResourceOptions}; -use metal::{MTLResourceUsage, MTLSamplerAddressMode, MTLSamplerMinMagFilter, MTLSize}; -use metal::{MTLStencilOperation, MTLStorageMode, MTLStoreAction, MTLTextureType, MTLTextureUsage}; -use metal::{MTLVertexFormat, MTLVertexStepFunction, MTLViewport, RenderCommandEncoder}; -use metal::{RenderCommandEncoderRef, RenderPassDescriptor, RenderPassDescriptorRef}; -use metal::{RenderPipelineColorAttachmentDescriptorRef, RenderPipelineDescriptor}; -use metal::{RenderPipelineReflection, RenderPipelineReflectionRef, RenderPipelineState}; -use metal::{SamplerDescriptor, SamplerState, StencilDescriptor, StructMemberRef, StructType}; -use metal::{StructTypeRef, TextureDescriptor, Texture, TextureRef, VertexAttribute}; -use metal::{VertexAttributeRef, VertexDescriptor, VertexDescriptorRef}; +use metal::{MTLArgument, MTLArgumentEncoder, MTLArgumentType, MTLBlendFactor, MTLBlendOperation}; +use metal::{MTLClearColor, MTLColorWriteMask, MTLCompareFunction, MTLComputePipelineState}; +use metal::{MTLDataType, MTLDevice, MTLIndexType, MTLLoadAction, MTLOrigin, MTLPixelFormat}; +use metal::{MTLPrimitiveType, MTLRegion, MTLRenderPipelineReflection, MTLRenderPipelineState}; +use metal::{MTLResourceOptions, MTLResourceUsage, MTLSamplerAddressMode, MTLSamplerMinMagFilter}; +use metal::{MTLSize, MTLStencilOperation, MTLStorageMode, MTLStoreAction, MTLTextureType}; +use metal::{MTLTextureUsage, MTLVertexFormat, MTLVertexStepFunction, MTLViewport}; +use metal::{RenderCommandEncoder, RenderCommandEncoderRef, RenderPassDescriptor}; +use metal::{RenderPassDescriptorRef, RenderPipelineColorAttachmentDescriptorRef}; +use metal::{RenderPipelineDescriptor, RenderPipelineReflection, RenderPipelineReflectionRef}; +use metal::{RenderPipelineState, SamplerDescriptor, SamplerState, StencilDescriptor}; +use metal::{StructMemberRef, StructType, StructTypeRef, TextureDescriptor, Texture, TextureRef}; +use metal::{VertexAttribute, VertexAttributeRef, VertexDescriptor, VertexDescriptorRef}; use objc::runtime::{Class, Object}; use pathfinder_geometry::rect::RectI; use pathfinder_geometry::vector::{Vector2I, vec2i}; -use pathfinder_gpu::{BlendFactor, BlendOp, BufferData, BufferTarget, BufferUploadMode, DepthFunc}; -use pathfinder_gpu::{Device, Primitive, RenderState, RenderTarget, ShaderKind, StencilFunc}; -use pathfinder_gpu::{TextureData, TextureDataRef, TextureFormat, TextureSamplingFlags}; -use pathfinder_gpu::{UniformData, VertexAttrClass, VertexAttrDescriptor, VertexAttrType}; +use pathfinder_gpu::{BlendFactor, BlendOp, BufferData, BufferTarget, BufferUploadMode}; +use pathfinder_gpu::{ComputeDimensions, ComputeState, DepthFunc, Device, ImageAccess, Primitive}; +use pathfinder_gpu::{ProgramKind, RenderState, RenderTarget, ShaderKind, StencilFunc, TextureData}; +use pathfinder_gpu::{TextureDataRef, TextureFormat, TextureSamplingFlags, UniformData}; +use pathfinder_gpu::{VertexAttrClass, VertexAttrDescriptor, VertexAttrType}; use pathfinder_resources::ResourceLoader; use pathfinder_simd::default::{F32x2, F32x4, I32x2}; use std::cell::{Cell, RefCell}; use std::mem; +use std::ops::Range; use std::ptr; use std::rc::Rc; use std::slice; use std::sync::{Arc, Condvar, Mutex, MutexGuard}; use std::time::{Duration, Instant}; -const FIRST_VERTEX_BUFFER_INDEX: u64 = 1; +const FIRST_VERTEX_BUFFER_INDEX: u64 = 16; pub struct MetalDevice { device: metal::Device, - layer: CoreAnimationLayer, - drawable: CoreAnimationDrawable, + main_color_texture: Texture, main_depth_stencil_texture: Texture, command_queue: CommandQueue, command_buffers: RefCell>, @@ -72,22 +76,30 @@ pub struct MetalDevice { next_timer_query_event_value: Cell, } -pub struct MetalProgram { - vertex: MetalShader, - fragment: MetalShader, +pub enum MetalProgram { + Raster(MetalRasterProgram), + Compute(MetalComputeProgram), +} + +pub struct MetalRasterProgram { + vertex_shader: MetalShader, + fragment_shader: MetalShader, +} + +pub struct MetalComputeProgram { + shader: MetalShader, + local_size: MTLSize, } #[derive(Clone)] pub struct MetalBuffer { buffer: Rc>>, + mode: BufferUploadMode, } impl MetalDevice { #[inline] - pub fn new(layer: &CoreAnimationLayerRef) -> MetalDevice { - let layer = layer.retain(); - let device = layer.device(); - let drawable = layer.next_drawable().unwrap().retain(); + pub unsafe fn new(device: metal::Device, texture: T) -> MetalDevice where T: IntoTexture { let command_queue = device.new_command_queue(); let samplers = (0..16).map(|sampling_flags_value| { @@ -122,17 +134,15 @@ impl MetalDevice { device.new_sampler(&sampler_descriptor) }).collect(); - let main_color_texture = drawable.texture(); - let framebuffer_size = vec2i(main_color_texture.width() as i32, - main_color_texture.height() as i32); + let texture = texture.into_texture(&device); + let framebuffer_size = vec2i(texture.width() as i32, texture.height() as i32); let main_depth_stencil_texture = device.create_depth_stencil_texture(framebuffer_size); let shared_event = device.new_shared_event(); MetalDevice { device, - layer, - drawable, + main_color_texture: texture, main_depth_stencil_texture, command_queue, command_buffers: RefCell::new(vec![]), @@ -143,11 +153,17 @@ impl MetalDevice { } } - pub fn present_drawable(&mut self) { - self.begin_commands(); - self.command_buffers.borrow_mut().last().unwrap().present_drawable(&self.drawable); - self.end_commands(); - self.drawable = self.layer.next_drawable().unwrap().retain(); + #[inline] + pub fn swap_texture(&mut self, new_texture: T) -> Texture where T: IntoTexture { + unsafe { + let new_texture = new_texture.into_texture(&self.device); + mem::replace(&mut self.main_color_texture, new_texture) + } + } + + #[inline] + pub fn metal_device(&self) -> metal::Device { + self.device.clone() } } @@ -157,13 +173,9 @@ pub struct MetalShader { #[allow(dead_code)] library: Library, function: Function, - uniforms: RefCell, -} - -enum ShaderUniforms { - Unknown, - NoUniforms, - Uniforms { encoder: ArgumentEncoder, struct_type: StructType } + #[allow(dead_code)] + name: String, + arguments: RefCell>, } pub struct MetalTexture { @@ -208,18 +220,38 @@ pub struct MetalUniform { name: String, } -#[derive(Clone, Copy)] -pub struct MetalUniformIndices { - vertex: Option, - fragment: Option, +#[derive(Clone)] +pub struct MetalStorageBuffer { + indices: RefCell>, + name: String, } -#[derive(Clone, Copy)] +#[derive(Clone, Copy, Debug)] +pub struct MetalUniformIndices(ProgramKind>); + +#[derive(Clone, Copy, Debug)] pub struct MetalUniformIndex { main: u64, sampler: Option, } +#[derive(Clone, Copy)] +pub struct MetalStorageBufferIndices(ProgramKind>); + +#[derive(Clone)] +pub struct MetalFence(Arc); + +struct MetalFenceInfo { + mutex: Mutex, + cond: Condvar, +} + +#[derive(Clone, Copy, PartialEq, Debug)] +enum MetalFenceStatus { + Pending, + Resolved, +} + pub struct MetalVertexArray { descriptor: VertexDescriptor, vertex_buffers: RefCell>, @@ -228,9 +260,11 @@ pub struct MetalVertexArray { impl Device for MetalDevice { type Buffer = MetalBuffer; + type Fence = MetalFence; type Framebuffer = MetalFramebuffer; type Program = MetalProgram; type Shader = MetalShader; + type StorageBuffer = MetalStorageBuffer; type Texture = MetalTexture; type TextureDataReceiver = MetalTextureDataReceiver; type TimerQuery = MetalTimerQuery; @@ -267,14 +301,19 @@ impl Device for MetalDevice { texture } - fn create_shader_from_source(&self, _: &str, source: &[u8], _: ShaderKind) -> MetalShader { + fn create_shader_from_source(&self, name: &str, source: &[u8], _: ShaderKind) -> MetalShader { let source = String::from_utf8(source.to_vec()).expect("Source wasn't valid UTF-8!"); let compile_options = CompileOptions::new(); let library = self.device.new_library_with_source(&source, &compile_options).unwrap(); let function = library.get_function("main0", None).unwrap(); - MetalShader { library, function, uniforms: RefCell::new(ShaderUniforms::Unknown) } + MetalShader { + library, + function, + name: name.to_owned(), + arguments: RefCell::new(None), + } } fn create_vertex_array(&self) -> MetalVertexArray { @@ -296,25 +335,52 @@ impl Device for MetalDevice { BufferTarget::Index => { *vertex_array.index_buffer.borrow_mut() = Some((*buffer).clone()) } + _ => panic!("Buffers bound to vertex arrays must be vertex or index buffers!"), } } fn create_program_from_shaders(&self, _: &dyn ResourceLoader, _: &str, - vertex_shader: MetalShader, - fragment_shader: MetalShader) + shaders: ProgramKind) -> MetalProgram { - MetalProgram { vertex: vertex_shader, fragment: fragment_shader } + match shaders { + ProgramKind::Raster { vertex: vertex_shader, fragment: fragment_shader } => { + MetalProgram::Raster(MetalRasterProgram { vertex_shader, fragment_shader }) + } + ProgramKind::Compute(shader) => { + let local_size = MTLSize { width: 0, height: 0, depth: 0 }; + MetalProgram::Compute(MetalComputeProgram { shader, local_size }) + } + } + } + + // FIXME(pcwalton): Is there a way to introspect the shader to find `gl_WorkGroupSize`? That + // would obviate the need for this function. + fn set_compute_program_local_size(&self, + program: &mut MetalProgram, + new_local_size: ComputeDimensions) { + match *program { + MetalProgram::Compute(MetalComputeProgram { ref mut local_size, .. }) => { + *local_size = new_local_size.to_metal_size() + } + _ => panic!("Program was not a compute program!"), + } } fn get_vertex_attr(&self, program: &MetalProgram, name: &str) -> Option { // TODO(pcwalton): Cache the function? - let attributes = program.vertex.function.real_vertex_attributes(); + let attributes = match *program { + MetalProgram::Raster(MetalRasterProgram { ref vertex_shader, .. }) => { + vertex_shader.function.real_vertex_attributes() + } + _ => unreachable!(), + }; for attribute_index in 0..attributes.len() { let attribute = attributes.object_at(attribute_index); let this_name = attribute.name().as_bytes(); if this_name[0] == b'a' && this_name[1..] == *name.as_bytes() { + //println!("found attribute: \"{}\"", name); return Some(attribute.retain()) } } @@ -325,6 +391,10 @@ impl Device for MetalDevice { MetalUniform { indices: RefCell::new(None), name: name.to_owned() } } + fn get_storage_buffer(&self, _: &Self::Program, name: &str, _: u32) -> MetalStorageBuffer { + MetalStorageBuffer { indices: RefCell::new(None), name: name.to_owned() } + } + fn configure_vertex_attr(&self, vertex_array: &MetalVertexArray, attr: &VertexAttribute, @@ -433,21 +503,15 @@ impl Device for MetalDevice { MetalFramebuffer(texture) } - fn create_buffer(&self) -> MetalBuffer { - MetalBuffer { buffer: Rc::new(RefCell::new(None)) } + fn create_buffer(&self, mode: BufferUploadMode) -> MetalBuffer { + MetalBuffer { buffer: Rc::new(RefCell::new(None)), mode } } fn allocate_buffer(&self, buffer: &MetalBuffer, data: BufferData, - _: BufferTarget, - mode: BufferUploadMode) { - let mut options = match mode { - BufferUploadMode::Static => MTLResourceOptions::CPUCacheModeWriteCombined, - BufferUploadMode::Dynamic => MTLResourceOptions::CPUCacheModeDefaultCache, - }; - options |= MTLResourceOptions::StorageModeManaged; - + _: BufferTarget) { + let options = buffer.mode.to_metal_resource_options(); match data { BufferData::Uninitialized(size) => { let size = (size * mem::size_of::()) as u64; @@ -464,6 +528,16 @@ impl Device for MetalDevice { } } + fn upload_to_buffer(&self, + buffer: &MetalBuffer, + start: usize, + data: &[T], + _: BufferTarget) { + let mut buffer = buffer.buffer.borrow_mut(); + let buffer = buffer.as_mut().unwrap(); + self.upload_to_metal_buffer(buffer, start, data) + } + #[inline] fn framebuffer_texture<'f>(&self, framebuffer: &'f MetalFramebuffer) -> &'f MetalTexture { &framebuffer.0 @@ -589,6 +663,59 @@ impl Device for MetalDevice { encoder.end_encoding(); } + fn dispatch_compute(&self, + size: ComputeDimensions, + compute_state: &ComputeState) { + let command_buffers = self.command_buffers.borrow(); + let command_buffer = command_buffers.last().unwrap(); + + let encoder = command_buffer.new_compute_command_encoder(); + + let program = match compute_state.program { + MetalProgram::Compute(ref compute_program) => compute_program, + _ => panic!("Compute render command must use a compute program!"), + }; + + let compute_pipeline_descriptor = ComputePipelineDescriptor::new(); + compute_pipeline_descriptor.set_compute_function(Some(&program.shader.function)); + + let compute_pipeline_state = unsafe { + if program.shader.arguments.borrow().is_none() { + // FIXME(pcwalton): Factor these raw Objective-C method calls out into a trait. + let mut reflection: *mut Object = ptr::null_mut(); + let reflection_options = MTLPipelineOption::ArgumentInfo | + MTLPipelineOption::BufferTypeInfo; + let mut error: *mut Object = ptr::null_mut(); + let raw_compute_pipeline_state: *mut MTLComputePipelineState = msg_send![ + self.device.as_ptr(), + newComputePipelineStateWithDescriptor:compute_pipeline_descriptor.as_ptr() + options:reflection_options + reflection:&mut reflection + error:&mut error]; + let compute_pipeline_state = + ComputePipelineState::from_ptr(raw_compute_pipeline_state); + *program.shader.arguments.borrow_mut() = + Some(ArgumentArray::from_ptr(msg_send![reflection, arguments])); + compute_pipeline_state + } else { + self.device + .new_compute_pipeline_state(&compute_pipeline_descriptor) + .expect("Failed to create compute pipeline state!") + } + }; + + self.set_compute_uniforms(&encoder, &compute_state); + encoder.set_compute_pipeline_state(&compute_pipeline_state); + + let local_size = match compute_state.program { + MetalProgram::Compute(MetalComputeProgram { ref local_size, .. }) => *local_size, + _ => panic!("Program was not a compute program!"), + }; + + encoder.dispatch_thread_groups(size.to_metal_size(), local_size); + encoder.end_encoding(); + } + fn create_timer_query(&self) -> MetalTimerQuery { let event_value = self.next_timer_query_event_value.get(); self.next_timer_query_event_value.set(event_value + 2); @@ -621,23 +748,19 @@ impl Device for MetalDevice { } fn begin_timer_query(&self, query: &MetalTimerQuery) { - /* self.command_buffers .borrow_mut() .last() .unwrap() .encode_signal_event(&self.shared_event, query.0.event_value); - */ } fn end_timer_query(&self, query: &MetalTimerQuery) { - /* self.command_buffers .borrow_mut() .last() .unwrap() .encode_signal_event(&self.shared_event, query.0.event_value + 1); - */ } fn try_recv_timer_query(&self, query: &MetalTimerQuery) -> Option { @@ -680,51 +803,146 @@ impl Device for MetalDevice { let suffix = match kind { ShaderKind::Vertex => 'v', ShaderKind::Fragment => 'f', + ShaderKind::Compute => 'c', }; let path = format!("shaders/metal/{}.{}s.metal", name, suffix); self.create_shader_from_source(name, &resources.slurp(&path).unwrap(), kind) } + + fn add_fence(&self) -> MetalFence { + let fence = MetalFence(Arc::new(MetalFenceInfo { + mutex: Mutex::new(MetalFenceStatus::Pending), + cond: Condvar::new(), + })); + let captured_fence = fence.clone(); + let block = ConcreteBlock::new(move |_| { + *captured_fence.0.mutex.lock().unwrap() = MetalFenceStatus::Resolved; + captured_fence.0.cond.notify_all(); + }); + self.command_buffers.borrow_mut().last().unwrap().add_completed_handler(block.copy()); + self.end_commands(); + self.begin_commands(); + fence + } + + fn wait_for_fence(&self, fence: &MetalFence) { + let mut guard = fence.0.mutex.lock().unwrap(); + while let MetalFenceStatus::Pending = *guard { + guard = fence.0.cond.wait(guard).unwrap(); + } + } } impl MetalDevice { fn get_uniform_index(&self, shader: &MetalShader, name: &str) -> Option { - let uniforms = shader.uniforms.borrow(); - let struct_type = match *uniforms { - ShaderUniforms::Unknown => panic!("get_uniform_index() called before reflection!"), - ShaderUniforms::NoUniforms => return None, - ShaderUniforms::Uniforms { ref struct_type, .. } => struct_type, + let uniforms = shader.arguments.borrow(); + let arguments = match *uniforms { + None => panic!("get_uniform_index() called before reflection!"), + Some(ref arguments) => arguments, }; - let main_member = match struct_type.member_from_name(&format!("u{}", name)) { - None => return None, - Some(main_member) => main_member, - }; - let main_index = main_member.argument_index(); - let sampler_index = match struct_type.member_from_name(&format!("u{}Smplr", name)) { + let (main_name, sampler_name) = (format!("u{}", name), format!("u{}Smplr", name)); + let (mut main_argument, mut sampler_argument) = (None, None); + for argument_index in 0..arguments.len() { + let argument = arguments.object_at(argument_index); + let argument_name = argument.name(); + if argument_name == &main_name { + main_argument = Some(argument.index()); + } else if argument_name == &sampler_name { + sampler_argument = Some(argument.index()); + } + } + let uniform_index = match main_argument { None => None, - Some(sampler_member) => Some(sampler_member.argument_index()), + Some(main) => Some(MetalUniformIndex { main, sampler: sampler_argument }), }; - Some(MetalUniformIndex { main: main_index, sampler: sampler_index }) + uniform_index + } + + fn get_storage_buffer_index(&self, shader: &MetalShader, name: &str) -> Option { + let uniforms = shader.arguments.borrow(); + let arguments = match *uniforms { + None => panic!("get_storage_buffer_index() called before reflection!"), + Some(ref arguments) => arguments, + }; + let main_name = format!("i{}", name); + let mut main_argument = None; + for argument_index in 0..arguments.len() { + let argument = arguments.object_at(argument_index); + match argument.type_() { + MTLArgumentType::Buffer => {} + _ => continue, + } + match argument.buffer_data_type() { + MTLDataType::Struct => {} + _ => continue, + } + let struct_type = argument.buffer_struct_type(); + if struct_type.member_from_name(&main_name).is_some() { + main_argument = Some(argument.index()); + } + } + let storage_buffer_index = match main_argument { + None => None, + Some(main) => Some(main), + }; + storage_buffer_index } fn populate_uniform_indices_if_necessary(&self, uniform: &MetalUniform, program: &MetalProgram) { - let mut indices = uniform.indices.borrow_mut(); if indices.is_some() { return; } - *indices = Some(MetalUniformIndices { - vertex: self.get_uniform_index(&program.vertex, &uniform.name), - fragment: self.get_uniform_index(&program.fragment, &uniform.name), - }); + *indices = match program { + MetalProgram::Raster(MetalRasterProgram { + ref vertex_shader, + ref fragment_shader, + }) => { + Some(MetalUniformIndices(ProgramKind::Raster { + vertex: self.get_uniform_index(vertex_shader, &uniform.name), + fragment: self.get_uniform_index(fragment_shader, &uniform.name), + })) + } + MetalProgram::Compute(MetalComputeProgram { ref shader, .. }) => { + let uniform_index = self.get_uniform_index(shader, &uniform.name); + Some(MetalUniformIndices(ProgramKind::Compute(uniform_index))) + } + } + } + + fn populate_storage_buffer_indices_if_necessary(&self, + storage_buffer: &MetalStorageBuffer, + program: &MetalProgram) { + let mut indices = storage_buffer.indices.borrow_mut(); + if indices.is_some() { + return; + } + + *indices = match program { + MetalProgram::Raster(MetalRasterProgram { + ref vertex_shader, + ref fragment_shader, + }) => { + Some(MetalStorageBufferIndices(ProgramKind::Raster { + vertex: self.get_storage_buffer_index(vertex_shader, &storage_buffer.name), + fragment: self.get_storage_buffer_index(fragment_shader, &storage_buffer.name), + })) + } + MetalProgram::Compute(MetalComputeProgram { ref shader, .. }) => { + let storage_buffer_index = self.get_storage_buffer_index(shader, + &storage_buffer.name); + Some(MetalStorageBufferIndices(ProgramKind::Compute(storage_buffer_index))) + } + } } fn render_target_color_texture(&self, render_target: &RenderTarget) -> Texture { match *render_target { - RenderTarget::Default {..} => self.drawable.texture().retain(), + RenderTarget::Default {..} => self.main_color_texture.retain(), RenderTarget::Framebuffer(framebuffer) => framebuffer.0.texture.retain(), } } @@ -771,13 +989,14 @@ impl MetalDevice { let encoder = command_buffer.new_render_command_encoder(&render_pass_descriptor).retain(); self.set_viewport(&encoder, &render_state.viewport); + let program = match render_state.program { + MetalProgram::Raster(ref raster_program) => raster_program, + _ => panic!("Raster render command must use a raster program!"), + }; + let render_pipeline_descriptor = RenderPipelineDescriptor::new(); - render_pipeline_descriptor.set_vertex_function(Some(&render_state.program - .vertex - .function)); - render_pipeline_descriptor.set_fragment_function(Some(&render_state.program - .fragment - .function)); + render_pipeline_descriptor.set_vertex_function(Some(&program.vertex_shader.function)); + render_pipeline_descriptor.set_fragment_function(Some(&program.fragment_shader.function)); render_pipeline_descriptor.set_vertex_descriptor(Some(&render_state.vertex_array .descriptor)); @@ -795,14 +1014,28 @@ impl MetalDevice { render_pipeline_descriptor.set_stencil_attachment_pixel_format(depth_stencil_format); } - let reflection_options = MTLPipelineOption::ArgumentInfo | - MTLPipelineOption::BufferTypeInfo; - let (render_pipeline_state, reflection) = - self.device.real_new_render_pipeline_state_with_reflection(&render_pipeline_descriptor, - reflection_options); - - self.populate_shader_uniforms_if_necessary(&render_state.program.vertex, &reflection); - self.populate_shader_uniforms_if_necessary(&render_state.program.fragment, &reflection); + let render_pipeline_state = if program.vertex_shader.arguments.borrow().is_none() || + program.fragment_shader.arguments.borrow().is_none() { + let reflection_options = MTLPipelineOption::ArgumentInfo | + MTLPipelineOption::BufferTypeInfo; + let (render_pipeline_state, reflection) = + self.device + .real_new_render_pipeline_state_with_reflection(&render_pipeline_descriptor, + reflection_options); + let mut vertex_arguments = program.vertex_shader.arguments.borrow_mut(); + let mut fragment_arguments = program.fragment_shader.arguments.borrow_mut(); + if vertex_arguments.is_none() { + *vertex_arguments = Some(reflection.real_vertex_arguments()); + } + if fragment_arguments.is_none() { + *fragment_arguments = Some(reflection.real_fragment_arguments()); + } + render_pipeline_state + } else { + self.device + .new_render_pipeline_state(&render_pipeline_descriptor) + .expect("Failed to create render pipeline state!") + }; for (vertex_buffer_index, vertex_buffer) in render_state.vertex_array .vertex_buffers @@ -815,98 +1048,108 @@ impl MetalDevice { .map(|buffer| buffer.as_ref()) .expect("Where's the vertex buffer?"); encoder.set_vertex_buffer(real_index, Some(buffer), 0); - encoder.use_resource(buffer, MTLResourceUsage::Read); } - self.set_uniforms(&encoder, render_state); + self.set_raster_uniforms(&encoder, render_state); encoder.set_render_pipeline_state(&render_pipeline_state); self.set_depth_stencil_state(&encoder, render_state); encoder } - fn populate_shader_uniforms_if_necessary(&self, - shader: &MetalShader, - reflection: &RenderPipelineReflectionRef) { - let mut uniforms = shader.uniforms.borrow_mut(); - match *uniforms { - ShaderUniforms::Unknown => {} - ShaderUniforms::NoUniforms | ShaderUniforms::Uniforms { .. } => return, - } - - let arguments = match shader.function.function_type() { - MTLFunctionType::Vertex => reflection.real_vertex_arguments(), - MTLFunctionType::Fragment => reflection.real_fragment_arguments(), - _ => panic!("Unexpected shader function type!"), + fn set_raster_uniforms(&self, + render_command_encoder: &RenderCommandEncoderRef, + render_state: &RenderState) { + let program = match render_state.program { + MetalProgram::Raster(ref raster_program) => raster_program, + _ => unreachable!(), }; - let mut has_descriptor_set = false; - for argument_index in 0..arguments.len() { - let argument = arguments.object_at(argument_index); - if argument.name() == "spvDescriptorSet0" { - has_descriptor_set = true; - break; - } - } - if !has_descriptor_set { - *uniforms = ShaderUniforms::NoUniforms; + let vertex_arguments = program.vertex_shader.arguments.borrow(); + let fragment_arguments = program.fragment_shader.arguments.borrow(); + if vertex_arguments.is_none() && fragment_arguments.is_none() { return; } - let (encoder, argument) = shader.function.new_argument_encoder_with_reflection(0); - match argument.buffer_data_type() { - MTLDataType::Struct => {} - data_type => { - panic!("Unexpected data type for argument buffer: {}!", data_type as u32) + let uniform_buffer = self.create_uniform_buffer(&render_state.uniforms); + for (&(uniform, ref uniform_data), buffer_range) in + render_state.uniforms.iter().zip(uniform_buffer.ranges.iter()) { + self.populate_uniform_indices_if_necessary(uniform, &render_state.program); + + let indices = uniform.indices.borrow_mut(); + let indices = indices.as_ref().unwrap(); + let (vertex_indices, fragment_indices) = match indices.0 { + ProgramKind::Raster { ref vertex, ref fragment } => (vertex, fragment), + _ => unreachable!(), + }; + + if let Some(vertex_index) = *vertex_indices { + self.set_vertex_uniform(vertex_index, + uniform_data, + &uniform_buffer.data, + buffer_range, + render_command_encoder, + render_state); + } + if let Some(fragment_index) = *fragment_indices { + self.set_fragment_uniform(fragment_index, + uniform_data, + &uniform_buffer.data, + buffer_range, + render_command_encoder, + render_state); } } - let struct_type = argument.buffer_struct_type().retain(); - *uniforms = ShaderUniforms::Uniforms { encoder, struct_type }; } - fn create_argument_buffer(&self, shader: &MetalShader) -> Option { - let uniforms = shader.uniforms.borrow(); - let encoder = match *uniforms { - ShaderUniforms::Unknown => unreachable!(), - ShaderUniforms::NoUniforms => return None, - ShaderUniforms::Uniforms { ref encoder, .. } => encoder, - }; + fn set_compute_uniforms(&self, + compute_command_encoder: &ComputeCommandEncoderRef, + compute_state: &ComputeState) { + let uniform_buffer = self.create_uniform_buffer(&compute_state.uniforms); + for (&(uniform, ref uniform_data), buffer_range) in + compute_state.uniforms.iter().zip(uniform_buffer.ranges.iter()) { + self.populate_uniform_indices_if_necessary(uniform, &compute_state.program); - let buffer_options = MTLResourceOptions::CPUCacheModeDefaultCache | - MTLResourceOptions::StorageModeManaged; - let buffer = self.device.new_buffer(encoder.encoded_length(), buffer_options); - encoder.set_argument_buffer(&buffer, 0); - Some(buffer) + let indices = uniform.indices.borrow_mut(); + let indices = indices.as_ref().unwrap(); + let indices = match indices.0 { + ProgramKind::Compute(ref indices) => indices, + _ => unreachable!(), + }; + + if let Some(index) = *indices { + self.set_compute_uniform(index, + uniform_data, + &uniform_buffer.data, + buffer_range, + compute_command_encoder, + compute_state); + } + } + + // Set storage buffers. + for &(storage_buffer_id, storage_buffer_binding) in compute_state.storage_buffers { + self.populate_storage_buffer_indices_if_necessary(storage_buffer_id, + &compute_state.program); + + let indices = storage_buffer_id.indices.borrow_mut(); + let indices = indices.as_ref().unwrap(); + let indices = match indices.0 { + ProgramKind::Compute(ref indices) => indices, + _ => unreachable!(), + }; + + if let Some(index) = *indices { + if let Some(ref buffer) = *storage_buffer_binding.buffer.borrow() { + compute_command_encoder.set_buffer(index, Some(buffer), 0); + } + } + + } } - fn set_uniforms(&self, - render_command_encoder: &RenderCommandEncoderRef, - render_state: &RenderState) { - let vertex_argument_buffer = self.create_argument_buffer(&render_state.program.vertex); - let fragment_argument_buffer = self.create_argument_buffer(&render_state.program.fragment); - - let vertex_uniforms = render_state.program.vertex.uniforms.borrow(); - let fragment_uniforms = render_state.program.fragment.uniforms.borrow(); - - let (mut have_vertex_uniforms, mut have_fragment_uniforms) = (false, false); - if let ShaderUniforms::Uniforms { .. } = *vertex_uniforms { - have_vertex_uniforms = true; - let vertex_argument_buffer = vertex_argument_buffer.as_ref().unwrap(); - render_command_encoder.use_resource(vertex_argument_buffer, MTLResourceUsage::Read); - render_command_encoder.set_vertex_buffer(0, Some(vertex_argument_buffer), 0); - } - if let ShaderUniforms::Uniforms { .. } = *fragment_uniforms { - have_fragment_uniforms = true; - let fragment_argument_buffer = fragment_argument_buffer.as_ref().unwrap(); - render_command_encoder.use_resource(fragment_argument_buffer, MTLResourceUsage::Read); - render_command_encoder.set_fragment_buffer(0, Some(fragment_argument_buffer), 0); - } - - if !have_vertex_uniforms && !have_fragment_uniforms { - return; - } - + fn create_uniform_buffer(&self, uniforms: &[(&MetalUniform, UniformData)]) -> UniformBuffer { let (mut uniform_buffer_data, mut uniform_buffer_ranges) = (vec![], vec![]); - for &(_, uniform_data) in render_state.uniforms.iter() { + for &(_, uniform_data) in uniforms.iter() { let start_index = uniform_buffer_data.len(); match uniform_data { UniformData::Float(value) => { @@ -953,92 +1196,135 @@ impl MetalDevice { uniform_buffer_data.write_f32::(vector.z()).unwrap(); uniform_buffer_data.write_f32::(vector.w()).unwrap(); } - UniformData::TextureUnit(_) => {} + UniformData::TextureUnit(_) | UniformData::ImageUnit(_) => {} } let end_index = uniform_buffer_data.len(); + while uniform_buffer_data.len() % 256 != 0 { + uniform_buffer_data.push(0); + } uniform_buffer_ranges.push(start_index..end_index); } - let buffer_options = MTLResourceOptions::CPUCacheModeWriteCombined | - MTLResourceOptions::StorageModeManaged; - let data_buffer = self.device - .new_buffer_with_data(uniform_buffer_data.as_ptr() as *const _, - uniform_buffer_data.len() as u64, - buffer_options); - - for (&(uniform, ref uniform_data), buffer_range) in - render_state.uniforms.iter().zip(uniform_buffer_ranges.iter()) { - self.populate_uniform_indices_if_necessary(uniform, &render_state.program); - let indices = uniform.indices.borrow_mut(); - let indices = indices.as_ref().unwrap(); - if let Some(vertex_index) = indices.vertex { - if let ShaderUniforms::Uniforms { - encoder: ref argument_encoder, - .. - } = *vertex_uniforms { - self.set_uniform(vertex_index, - argument_encoder, - uniform_data, - &data_buffer, - buffer_range.start as u64, - render_command_encoder, - render_state); - } - } - if let Some(fragment_index) = indices.fragment { - if let ShaderUniforms::Uniforms { - encoder: ref argument_encoder, - .. - } = *fragment_uniforms { - self.set_uniform(fragment_index, - argument_encoder, - uniform_data, - &data_buffer, - buffer_range.start as u64, - render_command_encoder, - render_state); - } - } - } - - render_command_encoder.use_resource(&data_buffer, MTLResourceUsage::Read); - - // Metal expects the data buffer to remain live. (Issue #199.) - // FIXME(pcwalton): When do we deallocate this? What are the expected - // lifetime semantics? - mem::forget(data_buffer); - - if let Some(vertex_argument_buffer) = vertex_argument_buffer { - let range = NSRange::new(0, vertex_argument_buffer.length()); - vertex_argument_buffer.did_modify_range(range); - } - if let Some(fragment_argument_buffer) = fragment_argument_buffer { - let range = NSRange::new(0, fragment_argument_buffer.length()); - fragment_argument_buffer.did_modify_range(range); + UniformBuffer { + data: uniform_buffer_data, + ranges: uniform_buffer_ranges, } } - fn set_uniform(&self, - argument_index: MetalUniformIndex, - argument_encoder: &ArgumentEncoder, - uniform_data: &UniformData, - buffer: &Buffer, - buffer_offset: u64, - render_command_encoder: &RenderCommandEncoderRef, - render_state: &RenderState) { + fn set_vertex_uniform(&self, + argument_index: MetalUniformIndex, + uniform_data: &UniformData, + buffer: &[u8], + buffer_range: &Range, + render_command_encoder: &RenderCommandEncoderRef, + render_state: &RenderState) { match *uniform_data { UniformData::TextureUnit(unit) => { let texture = render_state.textures[unit as usize]; - argument_encoder.set_texture(&texture.texture, argument_index.main); - let mut resource_usage = MTLResourceUsage::Read; - if let Some(sampler_index) = argument_index.sampler { - let sampler = &self.samplers[texture.sampling_flags.get().bits() as usize]; - argument_encoder.set_sampler_state(sampler, sampler_index); - resource_usage |= MTLResourceUsage::Sample; - } - render_command_encoder.use_resource(&texture.texture, resource_usage); + self.encode_vertex_texture_uniform(argument_index, + render_command_encoder, + texture); } - _ => argument_encoder.set_buffer(buffer, buffer_offset, argument_index.main), + UniformData::ImageUnit(unit) => { + let image = &render_state.images[unit as usize]; + render_command_encoder.set_vertex_texture(argument_index.main, + Some(&image.texture.texture)); + } + _ => { + render_command_encoder.set_vertex_bytes( + argument_index.main, + (buffer_range.end - buffer_range.start) as u64, + &buffer[buffer_range.start as usize] as *const u8 as *const _) + } + } + } + + fn set_fragment_uniform(&self, + argument_index: MetalUniformIndex, + uniform_data: &UniformData, + buffer: &[u8], + buffer_range: &Range, + render_command_encoder: &RenderCommandEncoderRef, + render_state: &RenderState) { + match *uniform_data { + UniformData::TextureUnit(unit) => { + let texture = render_state.textures[unit as usize]; + self.encode_fragment_texture_uniform(argument_index, + render_command_encoder, + texture); + } + UniformData::ImageUnit(unit) => { + let image = &render_state.images[unit as usize]; + render_command_encoder.set_fragment_texture(argument_index.main, + Some(&image.texture.texture)); + } + _ => { + render_command_encoder.set_fragment_bytes( + argument_index.main, + (buffer_range.end - buffer_range.start) as u64, + &buffer[buffer_range.start as usize] as *const u8 as *const _) + } + } + } + + fn set_compute_uniform(&self, + argument_index: MetalUniformIndex, + uniform_data: &UniformData, + buffer: &[u8], + buffer_range: &Range, + compute_command_encoder: &ComputeCommandEncoderRef, + compute_state: &ComputeState) { + match *uniform_data { + UniformData::TextureUnit(unit) => { + let texture = compute_state.textures[unit as usize]; + self.encode_compute_texture_uniform(argument_index, + compute_command_encoder, + texture); + } + UniformData::ImageUnit(unit) => { + let image = &compute_state.images[unit as usize]; + compute_command_encoder.set_texture(argument_index.main, + Some(&image.texture.texture)); + } + _ => { + compute_command_encoder.set_bytes( + argument_index.main, + (buffer_range.end - buffer_range.start) as u64, + &buffer[buffer_range.start as usize] as *const u8 as *const _) + } + } + } + + fn encode_vertex_texture_uniform(&self, + argument_index: MetalUniformIndex, + render_command_encoder: &RenderCommandEncoderRef, + texture: &MetalTexture) { + render_command_encoder.set_vertex_texture(argument_index.main, Some(&texture.texture)); + if let Some(sampler_index) = argument_index.sampler { + let sampler = &self.samplers[texture.sampling_flags.get().bits() as usize]; + render_command_encoder.set_vertex_sampler_state(sampler_index, Some(sampler)); + } + } + + fn encode_fragment_texture_uniform(&self, + argument_index: MetalUniformIndex, + render_command_encoder: &RenderCommandEncoderRef, + texture: &MetalTexture) { + render_command_encoder.set_fragment_texture(argument_index.main, Some(&texture.texture)); + if let Some(sampler_index) = argument_index.sampler { + let sampler = &self.samplers[texture.sampling_flags.get().bits() as usize]; + render_command_encoder.set_fragment_sampler_state(sampler_index, Some(sampler)); + } + } + + fn encode_compute_texture_uniform(&self, + argument_index: MetalUniformIndex, + compute_command_encoder: &ComputeCommandEncoderRef, + texture: &MetalTexture) { + compute_command_encoder.set_texture(argument_index.main, Some(&texture.texture)); + if let Some(sampler_index) = argument_index.sampler { + let sampler = &self.samplers[texture.sampling_flags.get().bits() as usize]; + compute_command_encoder.set_sampler_state(sampler_index, Some(sampler)); } } @@ -1185,18 +1471,28 @@ impl MetalDevice { } fn synchronize_texture(&self, texture: &Texture, block: RcBlock<(*mut Object,), ()>) { - unsafe { - let command_buffers = self.command_buffers.borrow(); - let command_buffer = command_buffers.last().unwrap(); - let encoder = command_buffer.new_blit_command_encoder(); - encoder.synchronize_resource(&texture); - let () = msg_send![*command_buffer, addCompletedHandler:&*block]; - encoder.end_encoding(); - } + let command_buffers = self.command_buffers.borrow(); + let command_buffer = command_buffers.last().unwrap(); + let encoder = command_buffer.new_blit_command_encoder(); + encoder.synchronize_resource(&texture); + command_buffer.add_completed_handler(block); + encoder.end_encoding(); self.end_commands(); self.begin_commands(); } + + fn upload_to_metal_buffer(&self, buffer: &Buffer, start: usize, data: &[T]) { + unsafe { + let start = (start * mem::size_of::()) as u64; + let size = (data.len() * mem::size_of::()) as u64; + assert!(start + size <= buffer.length()); + ptr::copy_nonoverlapping(data.as_ptr() as *const u8, + (buffer.contents() as *mut u8).offset(start as isize), + size as usize); + buffer.did_modify_range(NSRange::new(start, size)); + } + } } trait DeviceExtra { @@ -1216,6 +1512,58 @@ impl DeviceExtra for metal::Device { } } +// Helper types + +struct UniformBuffer { + data: Vec, + ranges: Vec>, +} + +// Miscellaneous extra public methods + +impl MetalTexture { + #[inline] + pub fn metal_texture(&self) -> Texture { + self.texture.clone() + } +} + +pub trait IntoTexture { + unsafe fn into_texture(self, metal_device: &metal::Device) -> Texture; +} + +impl IntoTexture for Texture { + #[inline] + unsafe fn into_texture(self, _: &metal::Device) -> Texture { + self + } +} + +impl IntoTexture for IOSurfaceRef { + #[inline] + unsafe fn into_texture(self, metal_device: &metal::Device) -> Texture { + let width = IOSurfaceGetWidth(self); + let height = IOSurfaceGetHeight(self); + + let descriptor = TextureDescriptor::new(); + descriptor.set_texture_type(MTLTextureType::D2); + descriptor.set_pixel_format(MTLPixelFormat::BGRA8Unorm); + descriptor.set_width(width as u64); + descriptor.set_height(height as u64); + descriptor.set_storage_mode(MTLStorageMode::Managed); + descriptor.set_usage(MTLTextureUsage::Unknown); + + msg_send![*metal_device, newTextureWithDescriptor:descriptor iosurface:self plane:0] + } +} + +impl<'a> IntoTexture for &'a CoreAnimationDrawableRef { + #[inline] + unsafe fn into_texture(self, _: &metal::Device) -> Texture { + self.texture().retain() + } +} + // Conversion helpers trait BlendFactorExt { @@ -1254,6 +1602,33 @@ impl BlendOpExt for BlendOp { } } +trait BufferUploadModeExt { + fn to_metal_resource_options(self) -> MTLResourceOptions; +} + +impl BufferUploadModeExt for BufferUploadMode { + #[inline] + fn to_metal_resource_options(self) -> MTLResourceOptions { + let mut options = match self { + BufferUploadMode::Static => MTLResourceOptions::CPUCacheModeWriteCombined, + BufferUploadMode::Dynamic => MTLResourceOptions::CPUCacheModeDefaultCache, + }; + options |= MTLResourceOptions::StorageModeManaged; + options + } +} + +trait ComputeDimensionsExt { + fn to_metal_size(self) -> MTLSize; +} + +impl ComputeDimensionsExt for ComputeDimensions { + #[inline] + fn to_metal_size(self) -> MTLSize { + MTLSize { width: self.x as u64, height: self.y as u64, depth: self.z as u64 } + } +} + trait DepthFuncExt { fn to_metal_compare_function(self) -> MTLCompareFunction; } @@ -1267,6 +1642,20 @@ impl DepthFuncExt for DepthFunc { } } +trait ImageAccessExt { + fn to_metal_resource_usage(self) -> MTLResourceUsage; +} + +impl ImageAccessExt for ImageAccess { + fn to_metal_resource_usage(self) -> MTLResourceUsage { + match self { + ImageAccess::Read => MTLResourceUsage::Read, + ImageAccess::Write => MTLResourceUsage::Write, + ImageAccess::ReadWrite => MTLResourceUsage::Read | MTLResourceUsage::Write, + } + } +} + trait PrimitiveExt { fn to_metal_primitive(self) -> MTLPrimitiveType; } @@ -1301,7 +1690,7 @@ impl UniformDataExt for UniformData { fn as_bytes(&self) -> Option<&[u8]> { unsafe { match *self { - UniformData::TextureUnit(_) => None, + UniformData::TextureUnit(_) | UniformData::ImageUnit(_) => None, UniformData::Float(ref data) => { Some(slice::from_raw_parts(data as *const f32 as *const u8, 4 * 1)) } @@ -1559,6 +1948,7 @@ impl CoreAnimationLayerExt for CoreAnimationLayer { trait CommandBufferExt { fn encode_signal_event(&self, event: &SharedEvent, value: u64); + fn add_completed_handler(&self, block: RcBlock<(*mut Object,), ()>); } impl CommandBufferExt for CommandBuffer { @@ -1567,6 +1957,12 @@ impl CommandBufferExt for CommandBuffer { msg_send![self.as_ptr(), encodeSignalEvent:event.0 value:value] } } + + fn add_completed_handler(&self, block: RcBlock<(*mut Object,), ()>) { + unsafe { + msg_send![self.as_ptr(), addCompletedHandler:&*block] + } + } } trait DeviceExt { @@ -1658,12 +2054,17 @@ impl RenderPipelineReflectionExt for RenderPipelineReflectionRef { trait StructMemberExt { fn argument_index(&self) -> u64; + fn pointer_type(&self) -> *mut Object; } impl StructMemberExt for StructMemberRef { fn argument_index(&self) -> u64 { unsafe { msg_send![self.as_ptr(), argumentIndex] } } + + fn pointer_type(&self) -> *mut Object { + unsafe { msg_send![self.as_ptr(), pointerType] } + } } // Memory management helpers @@ -1760,3 +2161,10 @@ struct BlockExtra { dtor: BlockExtraDtor, // 0x18 signature: *const *const i8, // 0x20 } + +// TODO(pcwalton): These should go upstream to `core-foundation-rs`. +#[link(name = "IOSurface", kind = "framework")] +extern { + fn IOSurfaceGetWidth(buffer: IOSurfaceRef) -> size_t; + fn IOSurfaceGetHeight(buffer: IOSurfaceRef) -> size_t; +} diff --git a/renderer/src/allocator.rs b/renderer/src/allocator.rs index 6ca303a8..929ef25d 100644 --- a/renderer/src/allocator.rs +++ b/renderer/src/allocator.rs @@ -50,6 +50,7 @@ enum TreeNode { } #[derive(Clone, Copy, PartialEq, Debug)] +#[allow(dead_code)] pub enum AllocationMode { Atlas, OwnPage, diff --git a/renderer/src/builder.rs b/renderer/src/builder.rs index 88cbbf0c..344438e8 100644 --- a/renderer/src/builder.rs +++ b/renderer/src/builder.rs @@ -29,7 +29,6 @@ use pathfinder_content::render_target::RenderTargetId; use pathfinder_geometry::line_segment::{LineSegment2F, LineSegmentU4, LineSegmentU8}; use pathfinder_geometry::rect::{RectF, RectI}; use pathfinder_geometry::transform2d::Transform2F; -use pathfinder_geometry::util; use pathfinder_geometry::vector::{Vector2F, Vector2I, vec2f, vec2i}; use pathfinder_gpu::TextureSamplingFlags; use pathfinder_simd::default::{F32x4, I32x4}; @@ -464,30 +463,6 @@ impl<'a, 'b> SceneBuilder<'a, 'b> { display_item_index: usize, tile_page: u16, } - - /* - // Create a new `DrawTiles` display item if we don't have one or if we have to break a - // batch due to blend mode or paint page. Note that every path with a blend mode that - // requires a readable framebuffer needs its own batch. - // - // TODO(pcwalton): If we really wanted to, we could use tile maps to avoid - // batch breaks in some cases… - - // Fetch the destination alpha tiles buffer. - let culled_alpha_tiles = match *culled_tiles.display_list.last_mut().unwrap() { - CulledDisplayItem::DrawTiles(TileBatch { tiles: ref mut culled_alpha_tiles, .. }) => { - culled_alpha_tiles - } - _ => unreachable!(), - }; - - for alpha_tile in alpha_tiles { - let alpha_tile_coords = alpha_tile.tile_position(); - if layer_z_buffer.test(alpha_tile_coords, current_depth) { - culled_alpha_tiles.push(*alpha_tile); - } - } - */ } fn pack_tiles(&mut self, culled_tiles: CulledTiles) { diff --git a/renderer/src/gpu/options.rs b/renderer/src/gpu/options.rs index 3d0244a2..e34e61ec 100644 --- a/renderer/src/gpu/options.rs +++ b/renderer/src/gpu/options.rs @@ -17,13 +17,11 @@ use pathfinder_gpu::Device; #[derive(Default)] pub struct RendererOptions { pub background_color: Option, + pub use_compute: bool, } #[derive(Clone)] -pub enum DestFramebuffer -where - D: Device, -{ +pub enum DestFramebuffer where D: Device { Default { viewport: RectI, window_size: Vector2I, diff --git a/renderer/src/gpu/renderer.rs b/renderer/src/gpu/renderer.rs index 8343937c..b0e1f8ff 100644 --- a/renderer/src/gpu/renderer.rs +++ b/renderer/src/gpu/renderer.rs @@ -32,7 +32,7 @@ use pathfinder_geometry::transform3d::Transform4F; use pathfinder_geometry::util; use pathfinder_geometry::vector::{Vector2F, Vector2I, Vector4F, vec2f, vec2i}; use pathfinder_gpu::{BlendFactor, BlendOp, BlendState, BufferData, BufferTarget, BufferUploadMode}; -use pathfinder_gpu::{ClearOps, DepthFunc, DepthState, Device, Primitive, RenderOptions}; +use pathfinder_gpu::{ClearOps, ComputeDimensions, ComputeState, DepthFunc, DepthState, Device, ImageAccess, ImageBinding, Primitive, RenderOptions}; use pathfinder_gpu::{RenderState, RenderTarget, StencilFunc, StencilState, TextureDataRef}; use pathfinder_gpu::{TextureFormat, UniformData}; use pathfinder_resources::ResourceLoader; @@ -54,7 +54,6 @@ pub(crate) const MASK_TILES_DOWN: u32 = 256; const SQRT_2_PI_INV: f32 = 0.3989422804014327; const TEXTURE_CACHE_SIZE: usize = 8; -const TIMER_QUERY_CACHE_SIZE: usize = 8; const TEXTURE_METADATA_ENTRIES_PER_ROW: i32 = 128; const TEXTURE_METADATA_TEXTURE_WIDTH: i32 = TEXTURE_METADATA_ENTRIES_PER_ROW * 4; @@ -92,10 +91,7 @@ const COMBINER_CTRL_COLOR_FILTER_SHIFT: i32 = 4; const COMBINER_CTRL_COLOR_COMBINE_SHIFT: i32 = 6; const COMBINER_CTRL_COMPOSITE_SHIFT: i32 = 8; -pub struct Renderer -where - D: Device, -{ +pub struct Renderer where D: Device { // Device pub device: D, @@ -107,61 +103,64 @@ where tile_program: TileProgram, tile_copy_program: CopyTileProgram, tile_clip_program: ClipTileProgram, - blit_vertex_array: BlitVertexArray, - tile_vertex_array: TileVertexArray, - tile_copy_vertex_array: CopyTileVertexArray, - tile_clip_vertex_array: ClipTileVertexArray, - tile_vertex_buffer: D::Buffer, + stencil_program: StencilProgram, + reprojection_program: ReprojectionProgram, quad_vertex_positions_buffer: D::Buffer, quad_vertex_indices_buffer: D::Buffer, - quads_vertex_indices_buffer: D::Buffer, - quads_vertex_indices_length: usize, - fill_vertex_array: FillVertexArray, - alpha_tile_pages: FxHashMap>, - dest_blend_framebuffer: D::Framebuffer, - intermediate_dest_framebuffer: D::Framebuffer, + next_fills: Vec, + fill_tile_map: Vec, texture_pages: Vec>>, render_targets: Vec, render_target_stack: Vec, - texture_metadata_texture: D::Texture, area_lut_texture: D::Texture, gamma_lut_texture: D::Texture, - // Stencil shader - stencil_program: StencilProgram, - stencil_vertex_array: StencilVertexArray, - - // Reprojection shader - reprojection_program: ReprojectionProgram, - reprojection_vertex_array: ReprojectionVertexArray, + // Frames + front_frame: Frame, + back_frame: Frame, + front_frame_fence: Option, // Rendering state - framebuffer_flags: FramebufferFlags, texture_cache: TextureCache, // Debug pub stats: RenderStats, current_cpu_build_time: Option, - current_timer: Option, + current_timer: Option>, pending_timers: VecDeque>, - free_timer_queries: Vec, + timer_query_cache: TimerQueryCache, pub debug_ui_presenter: DebugUIPresenter, // Extra info flags: RendererFlags, } -impl Renderer -where - D: Device, -{ +struct Frame where D: Device { + framebuffer_flags: FramebufferFlags, + blit_vertex_array: BlitVertexArray, + tile_vertex_array: TileVertexArray, + tile_vertex_buffer: D::Buffer, + fill_vertex_storage_allocator: FillVertexStorageAllocator, + quads_vertex_indices_buffer: D::Buffer, + quads_vertex_indices_length: usize, + alpha_tile_pages: FxHashMap>, + tile_copy_vertex_array: CopyTileVertexArray, + tile_clip_vertex_array: ClipTileVertexArray, + stencil_vertex_array: StencilVertexArray, + reprojection_vertex_array: ReprojectionVertexArray, + dest_blend_framebuffer: D::Framebuffer, + intermediate_dest_framebuffer: D::Framebuffer, + texture_metadata_texture: D::Texture, +} + +impl Renderer where D: Device { pub fn new(device: D, resources: &dyn ResourceLoader, dest_framebuffer: DestFramebuffer, options: RendererOptions) -> Renderer { let blit_program = BlitProgram::new(&device, resources); - let fill_program = FillProgram::new(&device, resources); + let fill_program = FillProgram::new(&device, resources, &options); let tile_program = TileProgram::new(&device, resources); let tile_copy_program = CopyTileProgram::new(&device, resources); let tile_clip_program = ClipTileProgram::new(&device, resources); @@ -171,79 +170,41 @@ where let area_lut_texture = device.create_texture_from_png(resources, "area-lut"); let gamma_lut_texture = device.create_texture_from_png(resources, "gamma-lut"); - let texture_metadata_texture = device.create_texture( - TextureFormat::RGBA16F, - Vector2I::new(TEXTURE_METADATA_TEXTURE_WIDTH, TEXTURE_METADATA_TEXTURE_HEIGHT)); - - let quad_vertex_positions_buffer = device.create_buffer(); - device.allocate_buffer( - &quad_vertex_positions_buffer, - BufferData::Memory(&QUAD_VERTEX_POSITIONS), - BufferTarget::Vertex, - BufferUploadMode::Static, - ); - let quad_vertex_indices_buffer = device.create_buffer(); - device.allocate_buffer( - &quad_vertex_indices_buffer, - BufferData::Memory(&QUAD_VERTEX_INDICES), - BufferTarget::Index, - BufferUploadMode::Static, - ); - let quads_vertex_indices_buffer = device.create_buffer(); - let tile_vertex_buffer = device.create_buffer(); - - let blit_vertex_array = BlitVertexArray::new( - &device, - &blit_program, - &quad_vertex_positions_buffer, - &quad_vertex_indices_buffer, - ); - let fill_vertex_array = FillVertexArray::new( - &device, - &fill_program, - &quad_vertex_positions_buffer, - &quad_vertex_indices_buffer, - ); - let tile_vertex_array = TileVertexArray::new( - &device, - &tile_program, - &tile_vertex_buffer, - &quad_vertex_positions_buffer, - &quad_vertex_indices_buffer, - ); - let tile_copy_vertex_array = CopyTileVertexArray::new( - &device, - &tile_copy_program, - &tile_vertex_buffer, - &quads_vertex_indices_buffer, - ); - let tile_clip_vertex_array = ClipTileVertexArray::new( - &device, - &tile_clip_program, - &quad_vertex_positions_buffer, - &quad_vertex_indices_buffer, - ); - let stencil_vertex_array = StencilVertexArray::new(&device, &stencil_program); - let reprojection_vertex_array = ReprojectionVertexArray::new( - &device, - &reprojection_program, - &quad_vertex_positions_buffer, - &quad_vertex_indices_buffer, - ); + let quad_vertex_positions_buffer = device.create_buffer(BufferUploadMode::Static); + device.allocate_buffer(&quad_vertex_positions_buffer, + BufferData::Memory(&QUAD_VERTEX_POSITIONS), + BufferTarget::Vertex); + let quad_vertex_indices_buffer = device.create_buffer(BufferUploadMode::Static); + device.allocate_buffer(&quad_vertex_indices_buffer, + BufferData::Memory(&QUAD_VERTEX_INDICES), + BufferTarget::Index); let window_size = dest_framebuffer.window_size(&device); - let dest_blend_texture = device.create_texture(TextureFormat::RGBA8, window_size); - let dest_blend_framebuffer = device.create_framebuffer(dest_blend_texture); - let intermediate_dest_texture = device.create_texture(TextureFormat::RGBA8, window_size); - let intermediate_dest_framebuffer = device.create_framebuffer(intermediate_dest_texture); - - let mut timer_queries = vec![]; - for _ in 0..TIMER_QUERY_CACHE_SIZE { - timer_queries.push(device.create_timer_query()); - } + let timer_query_cache = TimerQueryCache::new(&device); let debug_ui_presenter = DebugUIPresenter::new(&device, resources, window_size); + let front_frame = Frame::new(&device, + &blit_program, + &tile_program, + &tile_copy_program, + &tile_clip_program, + &reprojection_program, + &stencil_program, + &quad_vertex_positions_buffer, + &quad_vertex_indices_buffer, + window_size); + let back_frame = Frame::new(&device, + &blit_program, + &tile_program, + &tile_copy_program, + &tile_clip_program, + &reprojection_program, + &stencil_program, + &quad_vertex_positions_buffer, + &quad_vertex_indices_buffer, + window_size); + Renderer { device, @@ -254,41 +215,32 @@ where tile_program, tile_copy_program, tile_clip_program, - blit_vertex_array, - tile_vertex_array, - tile_copy_vertex_array, - tile_clip_vertex_array, - tile_vertex_buffer, quad_vertex_positions_buffer, quad_vertex_indices_buffer, - quads_vertex_indices_buffer, - quads_vertex_indices_length: 0, - fill_vertex_array, - alpha_tile_pages: FxHashMap::default(), - dest_blend_framebuffer, - intermediate_dest_framebuffer, + next_fills: vec![], + fill_tile_map: vec![-1; 256 * 256], texture_pages: vec![], render_targets: vec![], render_target_stack: vec![], + front_frame, + back_frame, + front_frame_fence: None, + area_lut_texture, gamma_lut_texture, - texture_metadata_texture, stencil_program, - stencil_vertex_array, reprojection_program, - reprojection_vertex_array, stats: RenderStats::default(), current_cpu_build_time: None, current_timer: None, pending_timers: VecDeque::new(), - free_timer_queries: timer_queries, + timer_query_cache, debug_ui_presenter, - framebuffer_flags: FramebufferFlags::empty(), texture_cache: TextureCache::new(), flags: RendererFlags::empty(), @@ -296,12 +248,13 @@ where } pub fn begin_scene(&mut self) { - self.framebuffer_flags = FramebufferFlags::empty(); - for alpha_tile_page in self.alpha_tile_pages.values_mut() { + self.back_frame.framebuffer_flags = FramebufferFlags::empty(); + for alpha_tile_page in self.back_frame.alpha_tile_pages.values_mut() { alpha_tile_page.must_preserve_framebuffer = false; } self.device.begin_commands(); + self.current_timer = Some(PendingTimer::new()); self.stats = RenderStats::default(); } @@ -325,7 +278,8 @@ where } RenderCommand::AddFills(ref fills) => self.add_fills(fills), RenderCommand::FlushFills => { - let page_indices: Vec<_> = self.alpha_tile_pages.keys().cloned().collect(); + let page_indices: Vec<_> = + self.back_frame.alpha_tile_pages.keys().cloned().collect(); for page_index in page_indices { self.draw_buffered_fills(page_index) } @@ -333,7 +287,7 @@ where RenderCommand::ClipTiles(ref batches) => { batches.iter().for_each(|batch| self.draw_clip_batch(batch)) } - RenderCommand::BeginTileDrawing => self.begin_tile_drawing(), + RenderCommand::BeginTileDrawing => {} RenderCommand::PushRenderTarget(render_target_id) => { self.push_render_target(render_target_id) } @@ -348,29 +302,30 @@ where batch.blend_mode, batch.filter) } - RenderCommand::Finish { cpu_build_time } => self.stats.cpu_build_time = cpu_build_time, - } - } - - fn begin_tile_drawing(&mut self) { - if let Some(timer_query) = self.allocate_timer_query() { - self.device.begin_timer_query(&timer_query); - self.current_timer = Some(timer_query); + RenderCommand::Finish { cpu_build_time } => { + self.stats.cpu_build_time = cpu_build_time; + } } } pub fn end_scene(&mut self) { self.blit_intermediate_dest_framebuffer_if_necessary(); + let old_front_frame_fence = self.front_frame_fence.take(); + self.front_frame_fence = Some(self.device.add_fence()); self.device.end_commands(); - if let Some(timer_query) = self.current_timer.take() { - self.device.end_timer_query(&timer_query); - self.pending_timers.push_back(PendingTimer { - gpu_timer_query: timer_query, - }); + self.back_frame.fill_vertex_storage_allocator.end_frame(); + if let Some(timer) = self.current_timer.take() { + self.pending_timers.push_back(timer); } self.current_cpu_build_time = None; + + if let Some(old_front_frame_fence) = old_front_frame_fence { + self.device.wait_for_fence(&old_front_frame_fence); + } + + mem::swap(&mut self.front_frame, &mut self.back_frame); } fn start_rendering(&mut self, @@ -397,11 +352,12 @@ where } pub fn shift_rendering_time(&mut self) -> Option { - if let Some(pending_timer) = self.pending_timers.pop_front() { - if let Some(gpu_time) = - self.device.try_recv_timer_query(&pending_timer.gpu_timer_query) { - self.free_timer_queries.push(pending_timer.gpu_timer_query); - return Some(RenderTime { gpu_time }); + if let Some(mut pending_timer) = self.pending_timers.pop_front() { + for old_query in pending_timer.poll(&self.device) { + self.timer_query_cache.free(old_query); + } + if let Some(gpu_time) = pending_timer.total_time() { + return Some(RenderTime { gpu_time }) } self.pending_timers.push_front(pending_timer); } @@ -531,7 +487,7 @@ where texels.push(f16::default()) } - let texture = &mut self.texture_metadata_texture; + let texture = &mut self.back_frame.texture_metadata_texture; let width = TEXTURE_METADATA_TEXTURE_WIDTH; let height = texels.len() as i32 / (4 * TEXTURE_METADATA_TEXTURE_WIDTH); let rect = RectI::new(Vector2I::zero(), Vector2I::new(width, height)); @@ -539,16 +495,15 @@ where } fn upload_tiles(&mut self, tiles: &[Tile]) { - self.device.allocate_buffer(&self.tile_vertex_buffer, + self.device.allocate_buffer(&self.back_frame.tile_vertex_buffer, BufferData::Memory(&tiles), - BufferTarget::Vertex, - BufferUploadMode::Dynamic); + BufferTarget::Vertex); self.ensure_index_buffer(tiles.len()); } fn ensure_index_buffer(&mut self, mut length: usize) { length = length.next_power_of_two(); - if self.quads_vertex_indices_length >= length { + if self.back_frame.quads_vertex_indices_length >= length { return; } @@ -561,14 +516,11 @@ where ]); } - self.device.allocate_buffer( - &self.quads_vertex_indices_buffer, - BufferData::Memory(&indices), - BufferTarget::Index, - BufferUploadMode::Static, - ); + self.device.allocate_buffer(&self.back_frame.quads_vertex_indices_buffer, + BufferData::Memory(&indices), + BufferTarget::Index); - self.quads_vertex_indices_length = length; + self.back_frame.quads_vertex_indices_length = length; } fn add_fills(&mut self, fill_batch: &[FillBatchEntry]) { @@ -578,26 +530,50 @@ where self.stats.fill_count += fill_batch.len(); + // Make sure we don't split batches across draw calls. + let mut pages_to_flush = vec![]; for fill_batch_entry in fill_batch { let page = fill_batch_entry.page; - if !self.alpha_tile_pages.contains_key(&page) { - self.alpha_tile_pages.insert(page, AlphaTilePage::new(&mut self.device)); + if !self.back_frame.alpha_tile_pages.contains_key(&page) { + let alpha_tile_page = AlphaTilePage::new(&mut self.device); + self.back_frame.alpha_tile_pages.insert(page, alpha_tile_page); } - if self.alpha_tile_pages[&page].buffered_fills.len() == MAX_FILLS_PER_BATCH { - self.draw_buffered_fills(page); + if self.back_frame + .alpha_tile_pages[&page] + .buffered_fills + .len() == MAX_FILLS_PER_BATCH { + pages_to_flush.push(page); } - self.alpha_tile_pages + self.back_frame + .alpha_tile_pages .get_mut(&page) .unwrap() .buffered_fills .push(fill_batch_entry.fill); } + + for page in pages_to_flush { + self.draw_buffered_fills(page); + } } fn draw_buffered_fills(&mut self, page: u16) { + match self.fill_program { + FillProgram::Raster(_) => self.draw_buffered_fills_via_raster(page), + FillProgram::Compute(_) => self.draw_buffered_fills_via_compute(page), + } + } + + fn draw_buffered_fills_via_raster(&mut self, page: u16) { + let fill_raster_program = match self.fill_program { + FillProgram::Raster(ref fill_raster_program) => fill_raster_program, + _ => unreachable!(), + }; + let mask_viewport = self.mask_viewport(); - let alpha_tile_page = self.alpha_tile_pages + let alpha_tile_page = self.back_frame + .alpha_tile_pages .get_mut(&page) .expect("Where's the alpha tile page?"); let buffered_fills = &mut alpha_tile_page.buffered_fills; @@ -605,33 +581,47 @@ where return; } - self.device.allocate_buffer( - &self.fill_vertex_array.vertex_buffer, - BufferData::Memory(&buffered_fills), - BufferTarget::Vertex, - BufferUploadMode::Dynamic, - ); + let fill_vertex_storage = self.back_frame + .fill_vertex_storage_allocator + .allocate(&self.device, + &self.fill_program, + &self.quad_vertex_positions_buffer, + &self.quad_vertex_indices_buffer); + + let fill_vertex_array = match fill_vertex_storage.auxiliary { + FillVertexStorageAuxiliary::Raster { ref vertex_array } => vertex_array, + _ => unreachable!(), + }; + + self.device.upload_to_buffer(&fill_vertex_storage.vertex_buffer, + 0, + &buffered_fills, + BufferTarget::Vertex); let mut clear_color = None; if !alpha_tile_page.must_preserve_framebuffer { clear_color = Some(ColorF::default()); }; + let timer_query = self.timer_query_cache.alloc(&self.device); + self.device.begin_timer_query(&timer_query); + debug_assert!(buffered_fills.len() <= u32::MAX as usize); self.device.draw_elements_instanced(6, buffered_fills.len() as u32, &RenderState { target: &RenderTarget::Framebuffer(&alpha_tile_page.framebuffer), - program: &self.fill_program.program, - vertex_array: &self.fill_vertex_array.vertex_array, + program: &fill_raster_program.program, + vertex_array: &fill_vertex_array.vertex_array, primitive: Primitive::Triangles, textures: &[&self.area_lut_texture], uniforms: &[ - (&self.fill_program.framebuffer_size_uniform, + (&fill_raster_program.framebuffer_size_uniform, UniformData::Vec2(F32x2::new(MASK_FRAMEBUFFER_WIDTH as f32, MASK_FRAMEBUFFER_HEIGHT as f32))), - (&self.fill_program.tile_size_uniform, + (&fill_raster_program.tile_size_uniform, UniformData::Vec2(F32x2::new(TILE_WIDTH as f32, TILE_HEIGHT as f32))), - (&self.fill_program.area_lut_uniform, UniformData::TextureUnit(0)), + (&fill_raster_program.area_lut_uniform, UniformData::TextureUnit(0)), ], + images: &[], viewport: mask_viewport, options: RenderOptions { blend: Some(BlendState { @@ -646,6 +636,102 @@ where }, }); + self.device.end_timer_query(&timer_query); + self.current_timer.as_mut().unwrap().fill_times.push(TimerFuture::new(timer_query)); + + alpha_tile_page.must_preserve_framebuffer = true; + buffered_fills.clear(); + } + + fn draw_buffered_fills_via_compute(&mut self, page: u16) { + let fill_compute_program = match self.fill_program { + FillProgram::Compute(ref fill_compute_program) => fill_compute_program, + _ => unreachable!(), + }; + + let alpha_tile_page = self.back_frame + .alpha_tile_pages + .get_mut(&page) + .expect("Where's the alpha tile page?"); + let buffered_fills = &mut alpha_tile_page.buffered_fills; + if buffered_fills.is_empty() { + return; + } + + let fill_vertex_storage = self.back_frame + .fill_vertex_storage_allocator + .allocate(&self.device, + &self.fill_program, + &self.quad_vertex_positions_buffer, + &self.quad_vertex_indices_buffer); + + let (tile_map_buffer, next_fills_buffer) = match fill_vertex_storage.auxiliary { + FillVertexStorageAuxiliary::Compute { ref tile_map_buffer, ref next_fills_buffer } => { + (tile_map_buffer, next_fills_buffer) + } + _ => unreachable!(), + }; + + // Initialize the tile map and fill linked list buffers. + self.fill_tile_map.iter_mut().for_each(|entry| *entry = -1); + while self.next_fills.len() < buffered_fills.len() { + self.next_fills.push(-1); + } + + // Create a linked list running through all our fills. + let (mut first_fill_tile, mut last_fill_tile) = (256 * 256, 0); + for (fill_index, fill) in buffered_fills.iter().enumerate() { + let fill_tile_index = fill.alpha_tile_index as usize; + self.next_fills[fill_index as usize] = self.fill_tile_map[fill_tile_index]; + self.fill_tile_map[fill_tile_index] = fill_index as i32; + first_fill_tile = first_fill_tile.min(fill_tile_index as u32); + last_fill_tile = last_fill_tile.max(fill_tile_index as u32); + } + let fill_tile_count = last_fill_tile - first_fill_tile + 1; + + self.device.upload_to_buffer(&fill_vertex_storage.vertex_buffer, + 0, + &buffered_fills, + BufferTarget::Storage); + self.device.upload_to_buffer(next_fills_buffer, + 0, + &self.next_fills, + BufferTarget::Storage); + self.device.upload_to_buffer(tile_map_buffer, + 0, + &self.fill_tile_map, + BufferTarget::Storage); + + let image_binding = ImageBinding { + texture: self.device.framebuffer_texture(&alpha_tile_page.framebuffer), + access: ImageAccess::Write, + }; + + let timer_query = self.timer_query_cache.alloc(&self.device); + self.device.begin_timer_query(&timer_query); + + debug_assert!(buffered_fills.len() <= u32::MAX as usize); + let dimensions = ComputeDimensions { x: 1, y: 1, z: fill_tile_count as u32 }; + self.device.dispatch_compute(dimensions, &ComputeState { + program: &fill_compute_program.program, + textures: &[&self.area_lut_texture], + images: &[image_binding], + uniforms: &[ + (&fill_compute_program.area_lut_uniform, UniformData::TextureUnit(0)), + (&fill_compute_program.dest_uniform, UniformData::ImageUnit(0)), + (&fill_compute_program.first_tile_index_uniform, + UniformData::Int(first_fill_tile as i32)), + ], + storage_buffers: &[ + (&fill_compute_program.fills_storage_buffer, &fill_vertex_storage.vertex_buffer), + (&fill_compute_program.next_fills_storage_buffer, next_fills_buffer), + (&fill_compute_program.fill_tile_map_storage_buffer, tile_map_buffer), + ], + }); + + self.device.end_timer_query(&timer_query); + self.current_timer.as_mut().unwrap().fill_times.push(TimerFuture::new(timer_query)); + alpha_tile_page.must_preserve_framebuffer = true; buffered_fills.clear(); } @@ -657,17 +743,17 @@ where let ClipBatchKey { dest_page, src_page, kind } = batch.key; - self.device.allocate_buffer(&self.tile_clip_vertex_array.vertex_buffer, + self.device.allocate_buffer(&self.back_frame.tile_clip_vertex_array.vertex_buffer, BufferData::Memory(&batch.clips), - BufferTarget::Vertex, - BufferUploadMode::Dynamic); + BufferTarget::Vertex); - if !self.alpha_tile_pages.contains_key(&dest_page) { - self.alpha_tile_pages.insert(dest_page, AlphaTilePage::new(&mut self.device)); + if !self.back_frame.alpha_tile_pages.contains_key(&dest_page) { + let alpha_tile_page = AlphaTilePage::new(&mut self.device); + self.back_frame.alpha_tile_pages.insert(dest_page, alpha_tile_page); } let mut clear_color = None; - if !self.alpha_tile_pages[&dest_page].must_preserve_framebuffer { + if !self.back_frame.alpha_tile_pages[&dest_page].must_preserve_framebuffer { clear_color = Some(ColorF::default()); }; @@ -686,18 +772,22 @@ where let mask_viewport = self.mask_viewport(); + let timer_query = self.timer_query_cache.alloc(&self.device); + self.device.begin_timer_query(&timer_query); + { - let dest_framebuffer = &self.alpha_tile_pages[&dest_page].framebuffer; - let src_framebuffer = &self.alpha_tile_pages[&src_page].framebuffer; + let dest_framebuffer = &self.back_frame.alpha_tile_pages[&dest_page].framebuffer; + let src_framebuffer = &self.back_frame.alpha_tile_pages[&src_page].framebuffer; let src_texture = self.device.framebuffer_texture(&src_framebuffer); debug_assert!(batch.clips.len() <= u32::MAX as usize); self.device.draw_elements_instanced(6, batch.clips.len() as u32, &RenderState { target: &RenderTarget::Framebuffer(dest_framebuffer), program: &self.tile_clip_program.program, - vertex_array: &self.tile_clip_vertex_array.vertex_array, + vertex_array: &self.back_frame.tile_clip_vertex_array.vertex_array, primitive: Primitive::Triangles, textures: &[src_texture], + images: &[], uniforms: &[(&self.tile_clip_program.src_uniform, UniformData::TextureUnit(0))], viewport: mask_viewport, options: RenderOptions { @@ -706,9 +796,16 @@ where ..RenderOptions::default() }, }); + + self.device.end_timer_query(&timer_query); + self.current_timer.as_mut().unwrap().fill_times.push(TimerFuture::new(timer_query)); } - self.alpha_tile_pages.get_mut(&dest_page).unwrap().must_preserve_framebuffer = true; + self.back_frame + .alpha_tile_pages + .get_mut(&dest_page) + .unwrap() + .must_preserve_framebuffer = true; } fn tile_transform(&self) -> Transform4F { @@ -733,7 +830,10 @@ where let clear_color = self.clear_color_for_draw_operation(); let draw_viewport = self.draw_viewport(); - let mut textures = vec![&self.texture_metadata_texture]; + let timer_query = self.timer_query_cache.alloc(&self.device); + self.device.begin_timer_query(&timer_query); + + let mut textures = vec![&self.back_frame.texture_metadata_texture]; let mut uniforms = vec![ (&self.tile_program.transform_uniform, UniformData::Mat4(self.tile_transform().to_columns())), @@ -750,10 +850,11 @@ where if needs_readable_framebuffer { uniforms.push((&self.tile_program.dest_texture_uniform, UniformData::TextureUnit(textures.len() as u32))); - textures.push(self.device.framebuffer_texture(&self.dest_blend_framebuffer)); + textures.push(self.device + .framebuffer_texture(&self.back_frame.dest_blend_framebuffer)); } - if let Some(alpha_tile_page) = self.alpha_tile_pages.get(&tile_page) { + if let Some(alpha_tile_page) = self.back_frame.alpha_tile_pages.get(&tile_page) { uniforms.push((&self.tile_program.mask_texture_0_uniform, UniformData::TextureUnit(textures.len() as u32))); textures.push(self.device.framebuffer_texture(&alpha_tile_page.framebuffer)); @@ -761,25 +862,32 @@ where // TODO(pcwalton): Refactor. let mut ctrl = 0; - if let Some(color_texture) = color_texture_0 { - let color_texture_page = self.texture_page(color_texture.page); - let color_texture_size = self.device.texture_size(color_texture_page).to_f32(); - self.device.set_texture_sampling_mode(color_texture_page, - color_texture.sampling_flags); - uniforms.push((&self.tile_program.color_texture_0_uniform, - UniformData::TextureUnit(textures.len() as u32))); - uniforms.push((&self.tile_program.color_texture_0_size_uniform, - UniformData::Vec2(color_texture_size.0))); - textures.push(color_texture_page); + match color_texture_0 { + Some(color_texture) => { + let color_texture_page = self.texture_page(color_texture.page); + let color_texture_size = self.device.texture_size(color_texture_page).to_f32(); + self.device.set_texture_sampling_mode(color_texture_page, + color_texture.sampling_flags); + uniforms.push((&self.tile_program.color_texture_0_uniform, + UniformData::TextureUnit(textures.len() as u32))); + uniforms.push((&self.tile_program.color_texture_0_size_uniform, + UniformData::Vec2(color_texture_size.0))); + textures.push(color_texture_page); - ctrl |= color_texture.composite_op.to_combine_mode() << - COMBINER_CTRL_COLOR_COMBINE_SHIFT; + ctrl |= color_texture.composite_op.to_combine_mode() << + COMBINER_CTRL_COLOR_COMBINE_SHIFT; + } + None => { + uniforms.push((&self.tile_program.color_texture_0_size_uniform, + UniformData::Vec2(F32x2::default()))); + } } ctrl |= blend_mode.to_composite_ctrl() << COMBINER_CTRL_COMPOSITE_SHIFT; + match filter { - Filter::None => {} + Filter::None => self.set_uniforms_for_no_filter(&mut uniforms), Filter::RadialGradient { line, radii, uv_origin } => { ctrl |= COMBINER_CTRL_FILTER_RADIAL_GRADIENT << COMBINER_CTRL_COLOR_FILTER_SHIFT; self.set_uniforms_for_radial_gradient_filter(&mut uniforms, line, radii, uv_origin) @@ -809,9 +917,10 @@ where self.device.draw_elements_instanced(6, tile_count, &RenderState { target: &self.draw_render_target(), program: &self.tile_program.program, - vertex_array: &self.tile_vertex_array.vertex_array, + vertex_array: &self.back_frame.tile_vertex_array.vertex_array, primitive: Primitive::Triangles, textures: &textures, + images: &[], uniforms: &uniforms, viewport: draw_viewport, options: RenderOptions { @@ -822,6 +931,9 @@ where }, }); + self.device.end_timer_query(&timer_query); + self.current_timer.as_mut().unwrap().tile_times.push(TimerFuture::new(timer_query)); + self.preserve_draw_framebuffer(); } @@ -849,11 +961,12 @@ where UniformData::Vec2(draw_viewport.size().to_f32().0))); self.device.draw_elements(tile_count * 6, &RenderState { - target: &RenderTarget::Framebuffer(&self.dest_blend_framebuffer), + target: &RenderTarget::Framebuffer(&self.back_frame.dest_blend_framebuffer), program: &self.tile_copy_program.program, - vertex_array: &self.tile_copy_vertex_array.vertex_array, + vertex_array: &self.back_frame.tile_copy_vertex_array.vertex_array, primitive: Primitive::Triangles, textures: &textures, + images: &[], uniforms: &uniforms, viewport: draw_viewport, options: RenderOptions { @@ -867,12 +980,9 @@ where } fn draw_stencil(&mut self, quad_positions: &[Vector4F]) { - self.device.allocate_buffer( - &self.stencil_vertex_array.vertex_buffer, - BufferData::Memory(quad_positions), - BufferTarget::Vertex, - BufferUploadMode::Dynamic, - ); + self.device.allocate_buffer(&self.back_frame.stencil_vertex_array.vertex_buffer, + BufferData::Memory(quad_positions), + BufferTarget::Vertex); // Create indices for a triangle fan. (This is OK because the clipped quad should always be // convex.) @@ -880,19 +990,17 @@ where for index in 1..(quad_positions.len() as u32 - 1) { indices.extend_from_slice(&[0, index as u32, index + 1]); } - self.device.allocate_buffer( - &self.stencil_vertex_array.index_buffer, - BufferData::Memory(&indices), - BufferTarget::Index, - BufferUploadMode::Dynamic, - ); + self.device.allocate_buffer(&self.back_frame.stencil_vertex_array.index_buffer, + BufferData::Memory(&indices), + BufferTarget::Index); self.device.draw_elements(indices.len() as u32, &RenderState { target: &self.draw_render_target(), program: &self.stencil_program.program, - vertex_array: &self.stencil_vertex_array.vertex_array, + vertex_array: &self.back_frame.stencil_vertex_array.vertex_array, primitive: Primitive::Triangles, textures: &[], + images: &[], uniforms: &[], viewport: self.draw_viewport(), options: RenderOptions { @@ -922,9 +1030,10 @@ where self.device.draw_elements(6, &RenderState { target: &self.draw_render_target(), program: &self.reprojection_program.program, - vertex_array: &self.reprojection_vertex_array.vertex_array, + vertex_array: &self.back_frame.reprojection_vertex_array.vertex_array, primitive: Primitive::Triangles, textures: &[texture], + images: &[], uniforms: &[ (&self.reprojection_program.old_transform_uniform, UniformData::from_transform_3d(old_transform)), @@ -953,7 +1062,7 @@ where } None => { if self.flags.contains(RendererFlags::INTERMEDIATE_DEST_FRAMEBUFFER_NEEDED) { - RenderTarget::Framebuffer(&self.intermediate_dest_framebuffer) + RenderTarget::Framebuffer(&self.back_frame.intermediate_dest_framebuffer) } else { match self.dest_framebuffer { DestFramebuffer::Default { .. } => RenderTarget::Default, @@ -974,6 +1083,15 @@ where self.render_target_stack.pop().expect("Render target stack underflow!"); } + fn set_uniforms_for_no_filter<'a>(&'a self, + uniforms: &mut Vec<(&'a D::Uniform, UniformData)>) { + uniforms.extend_from_slice(&[ + (&self.tile_program.filter_params_0_uniform, UniformData::Vec4(F32x4::default())), + (&self.tile_program.filter_params_1_uniform, UniformData::Vec4(F32x4::default())), + (&self.tile_program.filter_params_2_uniform, UniformData::Vec4(F32x4::default())), + ]); + } + fn set_uniforms_for_radial_gradient_filter<'a>( &'a self, uniforms: &mut Vec<(&'a D::Uniform, UniformData)>, @@ -985,6 +1103,7 @@ where UniformData::Vec4(line.from().0.concat_xy_xy(line.vector().0))), (&self.tile_program.filter_params_1_uniform, UniformData::Vec4(radii.concat_xy_xy(uv_origin.0))), + (&self.tile_program.filter_params_2_uniform, UniformData::Vec4(F32x4::default())), ]); } @@ -1042,6 +1161,7 @@ where UniformData::Vec4(src_offset.0.concat_xy_xy(F32x2::new(support, 0.0)))), (&self.tile_program.filter_params_1_uniform, UniformData::Vec4(F32x4::new(gauss_coeff_x, gauss_coeff_y, gauss_coeff_z, 0.0))), + (&self.tile_program.filter_params_2_uniform, UniformData::Vec4(F32x4::default())), ]); } @@ -1053,14 +1173,17 @@ where let main_viewport = self.main_viewport(); let uniforms = [(&self.blit_program.src_uniform, UniformData::TextureUnit(0))]; - let textures = [(self.device.framebuffer_texture(&self.intermediate_dest_framebuffer))]; + let textures = [ + (self.device.framebuffer_texture(&self.back_frame.intermediate_dest_framebuffer)) + ]; self.device.draw_elements(6, &RenderState { target: &RenderTarget::Default, program: &self.blit_program.program, - vertex_array: &self.blit_vertex_array.vertex_array, + vertex_array: &self.back_frame.blit_vertex_array.vertex_array, primitive: Primitive::Triangles, textures: &textures[..], + images: &[], uniforms: &uniforms[..], viewport: main_viewport, options: RenderOptions { @@ -1096,7 +1219,8 @@ where .must_preserve_contents } None => { - self.framebuffer_flags + self.back_frame + .framebuffer_flags .contains(FramebufferFlags::MUST_PRESERVE_DEST_FRAMEBUFFER_CONTENTS) } }; @@ -1120,7 +1244,8 @@ where .must_preserve_contents = true; } None => { - self.framebuffer_flags + self.back_frame + .framebuffer_flags .insert(FramebufferFlags::MUST_PRESERVE_DEST_FRAMEBUFFER_CONTENTS); } } @@ -1163,9 +1288,168 @@ where fn texture_page(&self, id: TexturePageId) -> &D::Texture { self.device.framebuffer_texture(&self.texture_page_framebuffer(id)) } +} - fn allocate_timer_query(&mut self) -> Option { - self.free_timer_queries.pop() +impl Frame where D: Device { + // FIXME(pcwalton): This signature shouldn't be so big. Make a struct. + fn new(device: &D, + blit_program: &BlitProgram, + tile_program: &TileProgram, + tile_copy_program: &CopyTileProgram, + tile_clip_program: &ClipTileProgram, + reprojection_program: &ReprojectionProgram, + stencil_program: &StencilProgram, + quad_vertex_positions_buffer: &D::Buffer, + quad_vertex_indices_buffer: &D::Buffer, + window_size: Vector2I) + -> Frame { + let tile_vertex_buffer = device.create_buffer(BufferUploadMode::Dynamic); + let quads_vertex_indices_buffer = device.create_buffer(BufferUploadMode::Dynamic); + + let blit_vertex_array = BlitVertexArray::new(device, + &blit_program, + &quad_vertex_positions_buffer, + &quad_vertex_indices_buffer); + let tile_vertex_array = TileVertexArray::new(device, + &tile_program, + &tile_vertex_buffer, + &quad_vertex_positions_buffer, + &quad_vertex_indices_buffer); + let tile_copy_vertex_array = CopyTileVertexArray::new(device, + &tile_copy_program, + &tile_vertex_buffer, + &quads_vertex_indices_buffer); + let tile_clip_vertex_array = ClipTileVertexArray::new(device, + &tile_clip_program, + &quad_vertex_positions_buffer, + &quad_vertex_indices_buffer); + let reprojection_vertex_array = ReprojectionVertexArray::new(device, + &reprojection_program, + &quad_vertex_positions_buffer, + &quad_vertex_indices_buffer); + let stencil_vertex_array = StencilVertexArray::new(device, &stencil_program); + + let fill_vertex_storage_allocator = FillVertexStorageAllocator::new(device); + + let texture_metadata_texture_size = vec2i(TEXTURE_METADATA_TEXTURE_WIDTH, + TEXTURE_METADATA_TEXTURE_HEIGHT); + let texture_metadata_texture = device.create_texture(TextureFormat::RGBA16F, + texture_metadata_texture_size); + + let intermediate_dest_texture = device.create_texture(TextureFormat::RGBA8, window_size); + let intermediate_dest_framebuffer = device.create_framebuffer(intermediate_dest_texture); + + let dest_blend_texture = device.create_texture(TextureFormat::RGBA8, window_size); + let dest_blend_framebuffer = device.create_framebuffer(dest_blend_texture); + + Frame { + blit_vertex_array, + tile_vertex_array, + tile_copy_vertex_array, + tile_clip_vertex_array, + reprojection_vertex_array, + stencil_vertex_array, + fill_vertex_storage_allocator, + tile_vertex_buffer, + quads_vertex_indices_buffer, + quads_vertex_indices_length: 0, + alpha_tile_pages: FxHashMap::default(), + texture_metadata_texture, + intermediate_dest_framebuffer, + dest_blend_framebuffer, + framebuffer_flags: FramebufferFlags::empty(), + } + } +} + +// Buffer management + +struct FillVertexStorageAllocator where D: Device { + free: Vec>, + in_use: Vec>, +} + +struct FillVertexStorage where D: Device { + vertex_buffer: D::Buffer, + auxiliary: FillVertexStorageAuxiliary, +} + +enum FillVertexStorageAuxiliary where D: Device { + Raster { vertex_array: FillVertexArray }, + Compute { + next_fills_buffer: D::Buffer, + tile_map_buffer: D::Buffer, + }, +} + +impl FillVertexStorageAllocator where D: Device { + fn new(_: &D) -> FillVertexStorageAllocator { + FillVertexStorageAllocator { free: vec![], in_use: vec![] } + } + + fn allocate(&mut self, + device: &D, + fill_program: &FillProgram, + quad_vertex_positions_buffer: &D::Buffer, + quad_vertex_indices_buffer: &D::Buffer) + -> &FillVertexStorage { + match self.free.pop() { + Some(storage) => self.in_use.push(storage), + None => { + self.in_use.push(FillVertexStorage::new(device, + fill_program, + quad_vertex_positions_buffer, + quad_vertex_indices_buffer)); + } + } + self.in_use.last().unwrap() + } + + fn end_frame(&mut self) { + self.free.extend(mem::replace(&mut self.in_use, vec![]).into_iter()) + } +} + +impl FillVertexStorage where D: Device { + fn new(device: &D, + fill_program: &FillProgram, + quad_vertex_positions_buffer: &D::Buffer, + quad_vertex_indices_buffer: &D::Buffer) + -> FillVertexStorage { + let vertex_buffer = device.create_buffer(BufferUploadMode::Dynamic); + // FIXME(pcwalton): * 2 is a hack; fix. + let vertex_buffer_data: BufferData = BufferData::Uninitialized(MAX_FILLS_PER_BATCH * 10); + device.allocate_buffer(&vertex_buffer, vertex_buffer_data, BufferTarget::Vertex); + + let auxiliary = match fill_program { + FillProgram::Raster(ref fill_raster_program) => { + FillVertexStorageAuxiliary::Raster { + vertex_array: FillVertexArray::new(device, + fill_raster_program, + &vertex_buffer, + quad_vertex_positions_buffer, + quad_vertex_indices_buffer), + } + } + FillProgram::Compute(_) => { + let next_fills_buffer = device.create_buffer(BufferUploadMode::Dynamic); + let tile_map_buffer = device.create_buffer(BufferUploadMode::Dynamic); + // FIXME(pcwalton): * 2 is a hack; fix. + let next_fills_buffer_data: BufferData = + BufferData::Uninitialized(MAX_FILLS_PER_BATCH * 10); + let tile_map_buffer_data: BufferData = + BufferData::Uninitialized(256 * 256); + device.allocate_buffer(&next_fills_buffer, + next_fills_buffer_data, + BufferTarget::Storage); + device.allocate_buffer(&tile_map_buffer, + tile_map_buffer_data, + BufferTarget::Storage); + FillVertexStorageAuxiliary::Compute { next_fills_buffer, tile_map_buffer } + } + }; + + FillVertexStorage { vertex_buffer, auxiliary } } } @@ -1206,9 +1490,81 @@ impl Div for RenderStats { } } -#[derive(Clone, Copy, Debug)] +struct TimerQueryCache where D: Device { + free_queries: Vec, +} + struct PendingTimer where D: Device { - gpu_timer_query: D::TimerQuery, + fill_times: Vec>, + tile_times: Vec>, +} + +enum TimerFuture where D: Device { + Pending(D::TimerQuery), + Resolved(Duration), +} + +impl TimerQueryCache where D: Device { + fn new(_: &D) -> TimerQueryCache { + TimerQueryCache { free_queries: vec![] } + } + + fn alloc(&mut self, device: &D) -> D::TimerQuery { + self.free_queries.pop().unwrap_or_else(|| device.create_timer_query()) + } + + fn free(&mut self, old_query: D::TimerQuery) { + self.free_queries.push(old_query); + } +} + +impl PendingTimer where D: Device { + fn new() -> PendingTimer { + PendingTimer { fill_times: vec![], tile_times: vec![] } + } + + fn poll(&mut self, device: &D) -> Vec { + let mut old_queries = vec![]; + for future in self.fill_times.iter_mut().chain(self.tile_times.iter_mut()) { + if let Some(old_query) = future.poll(device) { + old_queries.push(old_query) + } + } + old_queries + } + + fn total_time(&self) -> Option { + let mut total = Duration::default(); + for future in self.fill_times.iter().chain(self.tile_times.iter()) { + match *future { + TimerFuture::Pending(_) => return None, + TimerFuture::Resolved(time) => total += time, + } + } + Some(total) + } +} + +impl TimerFuture where D: Device { + fn new(query: D::TimerQuery) -> TimerFuture { + TimerFuture::Pending(query) + } + + fn poll(&mut self, device: &D) -> Option { + let duration = match *self { + TimerFuture::Pending(ref query) => device.try_recv_timer_query(query), + TimerFuture::Resolved(_) => None, + }; + match duration { + None => None, + Some(duration) => { + match mem::replace(self, TimerFuture::Resolved(duration)) { + TimerFuture::Resolved(_) => unreachable!(), + TimerFuture::Pending(old_query) => Some(old_query), + } + } + } + } } #[derive(Clone, Copy, Debug)] diff --git a/renderer/src/gpu/shaders.rs b/renderer/src/gpu/shaders.rs index 34dbfd41..ec03dc5a 100644 --- a/renderer/src/gpu/shaders.rs +++ b/renderer/src/gpu/shaders.rs @@ -8,8 +8,9 @@ // option. This file may not be copied, modified, or distributed // except according to those terms. -use crate::gpu_data::Fill; -use pathfinder_gpu::{BufferData, BufferTarget, BufferUploadMode, Device, VertexAttrClass}; +use crate::gpu::options::RendererOptions; +use crate::tiles::{TILE_HEIGHT, TILE_WIDTH}; +use pathfinder_gpu::{BufferTarget, BufferUploadMode, ComputeDimensions, Device, VertexAttrClass}; use pathfinder_gpu::{VertexAttrDescriptor, VertexAttrType}; use pathfinder_resources::ResourceLoader; @@ -49,12 +50,8 @@ impl BlitVertexArray where D: Device { } } -pub struct FillVertexArray -where - D: Device, -{ +pub struct FillVertexArray where D: Device { pub vertex_array: D::VertexArray, - pub vertex_buffer: D::Buffer, } impl FillVertexArray @@ -63,21 +60,13 @@ where { pub fn new( device: &D, - fill_program: &FillProgram, + fill_program: &FillRasterProgram, + vertex_buffer: &D::Buffer, quad_vertex_positions_buffer: &D::Buffer, quad_vertex_indices_buffer: &D::Buffer, ) -> FillVertexArray { let vertex_array = device.create_vertex_array(); - let vertex_buffer = device.create_buffer(); - let vertex_buffer_data: BufferData = BufferData::Uninitialized(MAX_FILLS_PER_BATCH); - device.allocate_buffer( - &vertex_buffer, - vertex_buffer_data, - BufferTarget::Vertex, - BufferUploadMode::Dynamic, - ); - let tess_coord_attr = device.get_vertex_attr(&fill_program.program, "TessCoord").unwrap(); let from_px_attr = device.get_vertex_attr(&fill_program.program, "FromPx").unwrap(); let to_px_attr = device.get_vertex_attr(&fill_program.program, "ToPx").unwrap(); @@ -96,30 +85,12 @@ where buffer_index: 0, }); device.bind_buffer(&vertex_array, &vertex_buffer, BufferTarget::Vertex); - device.configure_vertex_attr(&vertex_array, &from_px_attr, &VertexAttrDescriptor { - size: 1, - class: VertexAttrClass::Int, - attr_type: VertexAttrType::U8, - stride: FILL_INSTANCE_SIZE, - offset: 0, - divisor: 1, - buffer_index: 1, - }); - device.configure_vertex_attr(&vertex_array, &to_px_attr, &VertexAttrDescriptor { - size: 1, - class: VertexAttrClass::Int, - attr_type: VertexAttrType::U8, - stride: FILL_INSTANCE_SIZE, - offset: 1, - divisor: 1, - buffer_index: 1, - }); device.configure_vertex_attr(&vertex_array, &from_subpx_attr, &VertexAttrDescriptor { size: 2, class: VertexAttrClass::FloatNorm, attr_type: VertexAttrType::U8, stride: FILL_INSTANCE_SIZE, - offset: 2, + offset: 0, divisor: 1, buffer_index: 1, }); @@ -128,10 +99,28 @@ where class: VertexAttrClass::FloatNorm, attr_type: VertexAttrType::U8, stride: FILL_INSTANCE_SIZE, + offset: 2, + divisor: 1, + buffer_index: 1, + }); + device.configure_vertex_attr(&vertex_array, &from_px_attr, &VertexAttrDescriptor { + size: 1, + class: VertexAttrClass::Int, + attr_type: VertexAttrType::U8, + stride: FILL_INSTANCE_SIZE, offset: 4, divisor: 1, buffer_index: 1, }); + device.configure_vertex_attr(&vertex_array, &to_px_attr, &VertexAttrDescriptor { + size: 1, + class: VertexAttrClass::Int, + attr_type: VertexAttrType::U8, + stride: FILL_INSTANCE_SIZE, + offset: 5, + divisor: 1, + buffer_index: 1, + }); device.configure_vertex_attr(&vertex_array, &tile_index_attr, &VertexAttrDescriptor { size: 1, class: VertexAttrClass::Int, @@ -143,7 +132,7 @@ where }); device.bind_buffer(&vertex_array, quad_vertex_indices_buffer, BufferTarget::Index); - FillVertexArray { vertex_array, vertex_buffer } + FillVertexArray { vertex_array } } } @@ -277,7 +266,7 @@ impl ClipTileVertexArray where D: Device { quad_vertex_indices_buffer: &D::Buffer) -> ClipTileVertexArray { let vertex_array = device.create_vertex_array(); - let vertex_buffer = device.create_buffer(); + let vertex_buffer = device.create_buffer(BufferUploadMode::Dynamic); let tile_offset_attr = device.get_vertex_attr(&clip_tile_program.program, "TileOffset").unwrap(); @@ -340,32 +329,42 @@ pub struct BlitProgram where D: Device { impl BlitProgram where D: Device { pub fn new(device: &D, resources: &dyn ResourceLoader) -> BlitProgram { - let program = device.create_program(resources, "blit"); + let program = device.create_raster_program(resources, "blit"); let src_uniform = device.get_uniform(&program, "Src"); BlitProgram { program, src_uniform } } } -pub struct FillProgram -where - D: Device, -{ +pub enum FillProgram where D: Device { + Raster(FillRasterProgram), + Compute(FillComputeProgram), +} + +impl FillProgram where D: Device { + pub fn new(device: &D, resources: &dyn ResourceLoader, options: &RendererOptions) + -> FillProgram { + if options.use_compute { + FillProgram::Compute(FillComputeProgram::new(device, resources)) + } else { + FillProgram::Raster(FillRasterProgram::new(device, resources)) + } + } +} + +pub struct FillRasterProgram where D: Device { pub program: D::Program, pub framebuffer_size_uniform: D::Uniform, pub tile_size_uniform: D::Uniform, pub area_lut_uniform: D::Uniform, } -impl FillProgram -where - D: Device, -{ - pub fn new(device: &D, resources: &dyn ResourceLoader) -> FillProgram { - let program = device.create_program(resources, "fill"); +impl FillRasterProgram where D: Device { + pub fn new(device: &D, resources: &dyn ResourceLoader) -> FillRasterProgram { + let program = device.create_raster_program(resources, "fill"); let framebuffer_size_uniform = device.get_uniform(&program, "FramebufferSize"); let tile_size_uniform = device.get_uniform(&program, "TileSize"); let area_lut_uniform = device.get_uniform(&program, "AreaLUT"); - FillProgram { + FillRasterProgram { program, framebuffer_size_uniform, tile_size_uniform, @@ -374,6 +373,41 @@ where } } +pub struct FillComputeProgram where D: Device { + pub program: D::Program, + pub dest_uniform: D::Uniform, + pub area_lut_uniform: D::Uniform, + pub first_tile_index_uniform: D::Uniform, + pub fills_storage_buffer: D::StorageBuffer, + pub next_fills_storage_buffer: D::StorageBuffer, + pub fill_tile_map_storage_buffer: D::StorageBuffer, +} + +impl FillComputeProgram where D: Device { + pub fn new(device: &D, resources: &dyn ResourceLoader) -> FillComputeProgram { + let mut program = device.create_compute_program(resources, "fill"); + let local_size = ComputeDimensions { x: TILE_WIDTH, y: TILE_HEIGHT, z: 1 }; + device.set_compute_program_local_size(&mut program, local_size); + + let dest_uniform = device.get_uniform(&program, "Dest"); + let area_lut_uniform = device.get_uniform(&program, "AreaLUT"); + let first_tile_index_uniform = device.get_uniform(&program, "FirstTileIndex"); + let fills_storage_buffer = device.get_storage_buffer(&program, "Fills", 0); + let next_fills_storage_buffer = device.get_storage_buffer(&program, "NextFills", 1); + let fill_tile_map_storage_buffer = device.get_storage_buffer(&program, "FillTileMap", 2); + + FillComputeProgram { + program, + dest_uniform, + area_lut_uniform, + first_tile_index_uniform, + fills_storage_buffer, + next_fills_storage_buffer, + fill_tile_map_storage_buffer, + } + } +} + pub struct TileProgram where D: Device { pub program: D::Program, pub transform_uniform: D::Uniform, @@ -395,7 +429,7 @@ pub struct TileProgram where D: Device { impl TileProgram where D: Device { pub fn new(device: &D, resources: &dyn ResourceLoader) -> TileProgram { - let program = device.create_program(resources, "tile"); + let program = device.create_raster_program(resources, "tile"); let transform_uniform = device.get_uniform(&program, "Transform"); let tile_size_uniform = device.get_uniform(&program, "TileSize"); let texture_metadata_uniform = device.get_uniform(&program, "TextureMetadata"); @@ -442,7 +476,7 @@ pub struct CopyTileProgram where D: Device { impl CopyTileProgram where D: Device { pub fn new(device: &D, resources: &dyn ResourceLoader) -> CopyTileProgram { - let program = device.create_program(resources, "tile_copy"); + let program = device.create_raster_program(resources, "tile_copy"); let transform_uniform = device.get_uniform(&program, "Transform"); let tile_size_uniform = device.get_uniform(&program, "TileSize"); let framebuffer_size_uniform = device.get_uniform(&program, "FramebufferSize"); @@ -464,7 +498,7 @@ pub struct ClipTileProgram where D: Device { impl ClipTileProgram where D: Device { pub fn new(device: &D, resources: &dyn ResourceLoader) -> ClipTileProgram { - let program = device.create_program(resources, "tile_clip"); + let program = device.create_raster_program(resources, "tile_clip"); let src_uniform = device.get_uniform(&program, "Src"); ClipTileProgram { program, src_uniform } } @@ -482,7 +516,7 @@ where D: Device, { pub fn new(device: &D, resources: &dyn ResourceLoader) -> StencilProgram { - let program = device.create_program(resources, "stencil"); + let program = device.create_raster_program(resources, "stencil"); StencilProgram { program } } } @@ -502,7 +536,8 @@ where { pub fn new(device: &D, stencil_program: &StencilProgram) -> StencilVertexArray { let vertex_array = device.create_vertex_array(); - let (vertex_buffer, index_buffer) = (device.create_buffer(), device.create_buffer()); + let vertex_buffer = device.create_buffer(BufferUploadMode::Static); + let index_buffer = device.create_buffer(BufferUploadMode::Static); let position_attr = device.get_vertex_attr(&stencil_program.program, "Position").unwrap(); @@ -537,7 +572,7 @@ where D: Device, { pub fn new(device: &D, resources: &dyn ResourceLoader) -> ReprojectionProgram { - let program = device.create_program(resources, "reproject"); + let program = device.create_raster_program(resources, "reproject"); let old_transform_uniform = device.get_uniform(&program, "OldTransform"); let new_transform_uniform = device.get_uniform(&program, "NewTransform"); let texture_uniform = device.get_uniform(&program, "Texture"); diff --git a/renderer/src/gpu_data.rs b/renderer/src/gpu_data.rs index 0b7efa84..2282df5e 100644 --- a/renderer/src/gpu_data.rs +++ b/renderer/src/gpu_data.rs @@ -147,12 +147,11 @@ pub struct FillBatchEntry { pub page: u16, } -// FIXME(pcwalton): Move `subpx` before `px` and remove `repr(packed)`. #[derive(Clone, Copy, Debug, Default)] -#[repr(packed)] +#[repr(C)] pub struct Fill { - pub px: LineSegmentU4, pub subpx: LineSegmentU8, + pub px: LineSegmentU4, pub alpha_tile_index: u16, } diff --git a/resources/shaders/gl3/fill.fs.glsl b/resources/shaders/gl3/fill.fs.glsl index 79822d19..09a8069b 100644 --- a/resources/shaders/gl3/fill.fs.glsl +++ b/resources/shaders/gl3/fill.fs.glsl @@ -12,21 +12,24 @@ +#extension GL_GOOGLE_include_directive : enable + precision highp float; precision highp sampler2D; -uniform sampler2D uAreaLUT; -in vec2 vFrom; -in vec2 vTo; -out vec4 oFragColor; -void main(){ - vec2 from = vFrom, to = vTo; + + + + + +float computeCoverage(vec2 from, vec2 to, sampler2D areaLUT){ + vec2 left = from . x < to . x ? from : to, right = from . x < to . x ? to : from; @@ -40,6 +43,18 @@ void main(){ float dX = window . x - window . y; - oFragColor = vec4(texture(uAreaLUT, vec2(y + 8.0, abs(d * dX))/ 16.0). r * dX); + return texture(areaLUT, vec2(y + 8.0, abs(d * dX))/ 16.0). r * dX; +} + + +uniform sampler2D uAreaLUT; + +in vec2 vFrom; +in vec2 vTo; + +out vec4 oFragColor; + +void main(){ + oFragColor = vec4(computeCoverage(vFrom, vTo, uAreaLUT)); } diff --git a/resources/shaders/metal/blit.fs.metal b/resources/shaders/metal/blit.fs.metal index 5f3d2748..cb0ab32a 100644 --- a/resources/shaders/metal/blit.fs.metal +++ b/resources/shaders/metal/blit.fs.metal @@ -4,12 +4,6 @@ using namespace metal; -struct spvDescriptorSetBuffer0 -{ - texture2d uSrc [[id(0)]]; - sampler uSrcSmplr [[id(1)]]; -}; - struct main0_out { float4 oFragColor [[color(0)]]; @@ -20,10 +14,10 @@ struct main0_in float2 vTexCoord [[user(locn0)]]; }; -fragment main0_out main0(main0_in in [[stage_in]], constant spvDescriptorSetBuffer0& spvDescriptorSet0 [[buffer(0)]]) +fragment main0_out main0(main0_in in [[stage_in]], texture2d uSrc [[texture(0)]], sampler uSrcSmplr [[sampler(0)]]) { main0_out out = {}; - float4 color = spvDescriptorSet0.uSrc.sample(spvDescriptorSet0.uSrcSmplr, in.vTexCoord); + float4 color = uSrc.sample(uSrcSmplr, in.vTexCoord); out.oFragColor = float4(color.xyz * color.w, color.w); return out; } diff --git a/resources/shaders/metal/debug_solid.fs.metal b/resources/shaders/metal/debug_solid.fs.metal index 49f35a94..d8be9fb6 100644 --- a/resources/shaders/metal/debug_solid.fs.metal +++ b/resources/shaders/metal/debug_solid.fs.metal @@ -4,20 +4,15 @@ using namespace metal; -struct spvDescriptorSetBuffer0 -{ - constant float4* uColor [[id(0)]]; -}; - struct main0_out { float4 oFragColor [[color(0)]]; }; -fragment main0_out main0(constant spvDescriptorSetBuffer0& spvDescriptorSet0 [[buffer(0)]]) +fragment main0_out main0(constant float4& uColor [[buffer(0)]]) { main0_out out = {}; - out.oFragColor = float4((*spvDescriptorSet0.uColor).xyz, 1.0) * (*spvDescriptorSet0.uColor).w; + out.oFragColor = float4(uColor.xyz, 1.0) * uColor.w; return out; } diff --git a/resources/shaders/metal/debug_solid.vs.metal b/resources/shaders/metal/debug_solid.vs.metal index ec9d440f..286c4394 100644 --- a/resources/shaders/metal/debug_solid.vs.metal +++ b/resources/shaders/metal/debug_solid.vs.metal @@ -4,11 +4,6 @@ using namespace metal; -struct spvDescriptorSetBuffer0 -{ - constant float2* uFramebufferSize [[id(0)]]; -}; - struct main0_out { float4 gl_Position [[position]]; @@ -19,10 +14,10 @@ struct main0_in int2 aPosition [[attribute(0)]]; }; -vertex main0_out main0(main0_in in [[stage_in]], constant spvDescriptorSetBuffer0& spvDescriptorSet0 [[buffer(0)]]) +vertex main0_out main0(main0_in in [[stage_in]], constant float2& uFramebufferSize [[buffer(0)]]) { main0_out out = {}; - float2 position = ((float2(in.aPosition) / (*spvDescriptorSet0.uFramebufferSize)) * 2.0) - float2(1.0); + float2 position = ((float2(in.aPosition) / uFramebufferSize) * 2.0) - float2(1.0); out.gl_Position = float4(position.x, -position.y, 0.0, 1.0); return out; } diff --git a/resources/shaders/metal/debug_texture.fs.metal b/resources/shaders/metal/debug_texture.fs.metal index 80a811f3..8888110a 100644 --- a/resources/shaders/metal/debug_texture.fs.metal +++ b/resources/shaders/metal/debug_texture.fs.metal @@ -4,13 +4,6 @@ using namespace metal; -struct spvDescriptorSetBuffer0 -{ - texture2d uTexture [[id(0)]]; - sampler uTextureSmplr [[id(1)]]; - constant float4* uColor [[id(2)]]; -}; - struct main0_out { float4 oFragColor [[color(0)]]; @@ -21,11 +14,11 @@ struct main0_in float2 vTexCoord [[user(locn0)]]; }; -fragment main0_out main0(main0_in in [[stage_in]], constant spvDescriptorSetBuffer0& spvDescriptorSet0 [[buffer(0)]]) +fragment main0_out main0(main0_in in [[stage_in]], constant float4& uColor [[buffer(0)]], texture2d uTexture [[texture(0)]], sampler uTextureSmplr [[sampler(0)]]) { main0_out out = {}; - float alpha = spvDescriptorSet0.uTexture.sample(spvDescriptorSet0.uTextureSmplr, in.vTexCoord).x * (*spvDescriptorSet0.uColor).w; - out.oFragColor = float4((*spvDescriptorSet0.uColor).xyz, 1.0) * alpha; + float alpha = uTexture.sample(uTextureSmplr, in.vTexCoord).x * uColor.w; + out.oFragColor = float4(uColor.xyz, 1.0) * alpha; return out; } diff --git a/resources/shaders/metal/debug_texture.vs.metal b/resources/shaders/metal/debug_texture.vs.metal index 87a31507..1f824608 100644 --- a/resources/shaders/metal/debug_texture.vs.metal +++ b/resources/shaders/metal/debug_texture.vs.metal @@ -4,12 +4,6 @@ using namespace metal; -struct spvDescriptorSetBuffer0 -{ - constant float2* uTextureSize [[id(0)]]; - constant float2* uFramebufferSize [[id(1)]]; -}; - struct main0_out { float2 vTexCoord [[user(locn0)]]; @@ -22,11 +16,11 @@ struct main0_in int2 aTexCoord [[attribute(1)]]; }; -vertex main0_out main0(main0_in in [[stage_in]], constant spvDescriptorSetBuffer0& spvDescriptorSet0 [[buffer(0)]]) +vertex main0_out main0(main0_in in [[stage_in]], constant float2& uTextureSize [[buffer(0)]], constant float2& uFramebufferSize [[buffer(1)]]) { main0_out out = {}; - out.vTexCoord = float2(in.aTexCoord) / (*spvDescriptorSet0.uTextureSize); - float2 position = ((float2(in.aPosition) / (*spvDescriptorSet0.uFramebufferSize)) * 2.0) - float2(1.0); + out.vTexCoord = float2(in.aTexCoord) / uTextureSize; + float2 position = ((float2(in.aPosition) / uFramebufferSize) * 2.0) - float2(1.0); out.gl_Position = float4(position.x, -position.y, 0.0, 1.0); return out; } diff --git a/resources/shaders/metal/demo_ground.fs.metal b/resources/shaders/metal/demo_ground.fs.metal index 7d54eaf1..9b6d05ae 100644 --- a/resources/shaders/metal/demo_ground.fs.metal +++ b/resources/shaders/metal/demo_ground.fs.metal @@ -4,12 +4,6 @@ using namespace metal; -struct spvDescriptorSetBuffer0 -{ - constant float4* uGridlineColor [[id(0)]]; - constant float4* uGroundColor [[id(1)]]; -}; - struct main0_out { float4 oFragColor [[color(0)]]; @@ -20,11 +14,11 @@ struct main0_in float2 vTexCoord [[user(locn0)]]; }; -fragment main0_out main0(main0_in in [[stage_in]], constant spvDescriptorSetBuffer0& spvDescriptorSet0 [[buffer(0)]]) +fragment main0_out main0(main0_in in [[stage_in]], constant float4& uGridlineColor [[buffer(0)]], constant float4& uGroundColor [[buffer(1)]]) { main0_out out = {}; float2 texCoordPx = fract(in.vTexCoord) / fwidth(in.vTexCoord); - out.oFragColor = select((*spvDescriptorSet0.uGroundColor), (*spvDescriptorSet0.uGridlineColor), bool4(any(texCoordPx <= float2(1.0)))); + out.oFragColor = select(uGroundColor, uGridlineColor, bool4(any(texCoordPx <= float2(1.0)))); return out; } diff --git a/resources/shaders/metal/demo_ground.vs.metal b/resources/shaders/metal/demo_ground.vs.metal index 997185f5..2ec42d00 100644 --- a/resources/shaders/metal/demo_ground.vs.metal +++ b/resources/shaders/metal/demo_ground.vs.metal @@ -4,12 +4,6 @@ using namespace metal; -struct spvDescriptorSetBuffer0 -{ - constant int* uGridlineCount [[id(0)]]; - constant float4x4* uTransform [[id(1)]]; -}; - struct main0_out { float2 vTexCoord [[user(locn0)]]; @@ -21,11 +15,11 @@ struct main0_in int2 aPosition [[attribute(0)]]; }; -vertex main0_out main0(main0_in in [[stage_in]], constant spvDescriptorSetBuffer0& spvDescriptorSet0 [[buffer(0)]]) +vertex main0_out main0(main0_in in [[stage_in]], constant int& uGridlineCount [[buffer(0)]], constant float4x4& uTransform [[buffer(1)]]) { main0_out out = {}; - out.vTexCoord = float2(in.aPosition * int2((*spvDescriptorSet0.uGridlineCount))); - out.gl_Position = (*spvDescriptorSet0.uTransform) * float4(int4(in.aPosition.x, 0, in.aPosition.y, 1)); + out.vTexCoord = float2(in.aPosition * int2(uGridlineCount)); + out.gl_Position = uTransform * float4(int4(in.aPosition.x, 0, in.aPosition.y, 1)); return out; } diff --git a/resources/shaders/metal/fill.cs.metal b/resources/shaders/metal/fill.cs.metal new file mode 100644 index 00000000..e7b20b48 --- /dev/null +++ b/resources/shaders/metal/fill.cs.metal @@ -0,0 +1,67 @@ +// Automatically generated from files in pathfinder/shaders/. Do not edit! +#pragma clang diagnostic ignored "-Wmissing-prototypes" + +#include +#include + +using namespace metal; + +struct bFillTileMap +{ + int iFillTileMap[1]; +}; + +struct bFills +{ + uint2 iFills[1]; +}; + +struct bNextFills +{ + int iNextFills[1]; +}; + +constant uint3 gl_WorkGroupSize [[maybe_unused]] = uint3(16u, 16u, 1u); + +static inline __attribute__((always_inline)) +float computeCoverage(thread const float2& from, thread const float2& to, thread const texture2d areaLUT, thread const sampler areaLUTSmplr) +{ + float2 left = select(to, from, bool2(from.x < to.x)); + float2 right = select(from, to, bool2(from.x < to.x)); + float2 window = fast::clamp(float2(from.x, to.x), float2(-0.5), float2(0.5)); + float offset = mix(window.x, window.y, 0.5) - left.x; + float t = offset / (right.x - left.x); + float y = mix(left.y, right.y, t); + float d = (right.y - left.y) / (right.x - left.x); + float dX = window.x - window.y; + return areaLUT.sample(areaLUTSmplr, (float2(y + 8.0, abs(d * dX)) / float2(16.0)), level(0.0)).x * dX; +} + +kernel void main0(constant int& uFirstTileIndex [[buffer(0)]], const device bFillTileMap& _165 [[buffer(1)]], const device bFills& _186 [[buffer(2)]], const device bNextFills& _269 [[buffer(3)]], texture2d uAreaLUT [[texture(0)]], texture2d uDest [[texture(1)]], sampler uAreaLUTSmplr [[sampler(0)]], uint3 gl_LocalInvocationID [[thread_position_in_threadgroup]], uint3 gl_WorkGroupID [[threadgroup_position_in_grid]]) +{ + int2 tileSubCoord = int2(gl_LocalInvocationID.xy); + uint tileIndexOffset = gl_WorkGroupID.z; + uint tileIndex = tileIndexOffset + uint(uFirstTileIndex); + int2 tileOrigin = int2(int(tileIndex & 255u), int((tileIndex >> 8u) & 255u)) * int2(16); + int2 destCoord = tileOrigin + tileSubCoord; + int fillIndex = _165.iFillTileMap[tileIndex]; + if (fillIndex < 0) + { + return; + } + float coverage = 0.0; + do + { + uint2 fill = _186.iFills[fillIndex]; + float2 from = float2(float(fill.y & 15u), float((fill.y >> 4u) & 15u)) + (float2(float(fill.x & 255u), float((fill.x >> 8u) & 255u)) / float2(256.0)); + float2 to = float2(float((fill.y >> 8u) & 15u), float((fill.y >> 12u) & 15u)) + (float2(float((fill.x >> 16u) & 255u), float((fill.x >> 24u) & 255u)) / float2(256.0)); + from -= (float2(tileSubCoord) + float2(0.5)); + to -= (float2(tileSubCoord) + float2(0.5)); + float2 param = from; + float2 param_1 = to; + coverage += computeCoverage(param, param_1, uAreaLUT, uAreaLUTSmplr); + fillIndex = _269.iNextFills[fillIndex]; + } while (fillIndex >= 0); + uDest.write(float4(coverage), uint2(destCoord)); +} + diff --git a/resources/shaders/metal/fill.fs.metal b/resources/shaders/metal/fill.fs.metal index 04b844ad..28fd8afa 100644 --- a/resources/shaders/metal/fill.fs.metal +++ b/resources/shaders/metal/fill.fs.metal @@ -1,15 +1,11 @@ // Automatically generated from files in pathfinder/shaders/. Do not edit! +#pragma clang diagnostic ignored "-Wmissing-prototypes" + #include #include using namespace metal; -struct spvDescriptorSetBuffer0 -{ - texture2d uAreaLUT [[id(0)]]; - sampler uAreaLUTSmplr [[id(1)]]; -}; - struct main0_out { float4 oFragColor [[color(0)]]; @@ -21,11 +17,9 @@ struct main0_in float2 vTo [[user(locn1)]]; }; -fragment main0_out main0(main0_in in [[stage_in]], constant spvDescriptorSetBuffer0& spvDescriptorSet0 [[buffer(0)]]) +static inline __attribute__((always_inline)) +float computeCoverage(thread const float2& from, thread const float2& to, thread const texture2d areaLUT, thread const sampler areaLUTSmplr) { - main0_out out = {}; - float2 from = in.vFrom; - float2 to = in.vTo; float2 left = select(to, from, bool2(from.x < to.x)); float2 right = select(from, to, bool2(from.x < to.x)); float2 window = fast::clamp(float2(from.x, to.x), float2(-0.5), float2(0.5)); @@ -34,7 +28,15 @@ fragment main0_out main0(main0_in in [[stage_in]], constant spvDescriptorSetBuff float y = mix(left.y, right.y, t); float d = (right.y - left.y) / (right.x - left.x); float dX = window.x - window.y; - out.oFragColor = float4(spvDescriptorSet0.uAreaLUT.sample(spvDescriptorSet0.uAreaLUTSmplr, (float2(y + 8.0, abs(d * dX)) / float2(16.0))).x * dX); + return areaLUT.sample(areaLUTSmplr, (float2(y + 8.0, abs(d * dX)) / float2(16.0))).x * dX; +} + +fragment main0_out main0(main0_in in [[stage_in]], texture2d uAreaLUT [[texture(0)]], sampler uAreaLUTSmplr [[sampler(0)]]) +{ + main0_out out = {}; + float2 param = in.vFrom; + float2 param_1 = in.vTo; + out.oFragColor = float4(computeCoverage(param, param_1, uAreaLUT, uAreaLUTSmplr)); return out; } diff --git a/resources/shaders/metal/fill.vs.metal b/resources/shaders/metal/fill.vs.metal index 87f77d2a..13c7b2d9 100644 --- a/resources/shaders/metal/fill.vs.metal +++ b/resources/shaders/metal/fill.vs.metal @@ -6,12 +6,6 @@ using namespace metal; -struct spvDescriptorSetBuffer0 -{ - constant float2* uTileSize [[id(0)]]; - constant float2* uFramebufferSize [[id(1)]]; -}; - struct main0_out { float2 vFrom [[user(locn0)]]; @@ -29,6 +23,7 @@ struct main0_in uint aTileIndex [[attribute(5)]]; }; +static inline __attribute__((always_inline)) float2 computeTileOffset(thread const uint& tileIndex, thread const float& stencilTextureWidth, thread float2 uTileSize) { uint tilesPerRow = uint(stencilTextureWidth / uTileSize.x); @@ -36,12 +31,12 @@ float2 computeTileOffset(thread const uint& tileIndex, thread const float& stenc return float2(tileOffset) * uTileSize; } -vertex main0_out main0(main0_in in [[stage_in]], constant spvDescriptorSetBuffer0& spvDescriptorSet0 [[buffer(0)]]) +vertex main0_out main0(main0_in in [[stage_in]], constant float2& uTileSize [[buffer(0)]], constant float2& uFramebufferSize [[buffer(1)]]) { main0_out out = {}; uint param = in.aTileIndex; - float param_1 = (*spvDescriptorSet0.uFramebufferSize).x; - float2 tileOrigin = computeTileOffset(param, param_1, (*spvDescriptorSet0.uTileSize)); + float param_1 = uFramebufferSize.x; + float2 tileOrigin = computeTileOffset(param, param_1, uTileSize); float2 from = float2(float(in.aFromPx & 15u), float(in.aFromPx >> 4u)) + in.aFromSubpx; float2 to = float2(float(in.aToPx & 15u), float(in.aToPx >> 4u)) + in.aToSubpx; float2 position; @@ -59,11 +54,11 @@ vertex main0_out main0(main0_in in [[stage_in]], constant spvDescriptorSetBuffer } else { - position.y = (*spvDescriptorSet0.uTileSize).y; + position.y = uTileSize.y; } out.vFrom = from - position; out.vTo = to - position; - float2 globalPosition = (((tileOrigin + position) / (*spvDescriptorSet0.uFramebufferSize)) * 2.0) - float2(1.0); + float2 globalPosition = (((tileOrigin + position) / uFramebufferSize) * 2.0) - float2(1.0); globalPosition.y = -globalPosition.y; out.gl_Position = float4(globalPosition, 0.0, 1.0); return out; diff --git a/resources/shaders/metal/reproject.fs.metal b/resources/shaders/metal/reproject.fs.metal index be9ff4dd..61f14046 100644 --- a/resources/shaders/metal/reproject.fs.metal +++ b/resources/shaders/metal/reproject.fs.metal @@ -4,13 +4,6 @@ using namespace metal; -struct spvDescriptorSetBuffer0 -{ - constant float4x4* uOldTransform [[id(0)]]; - texture2d uTexture [[id(1)]]; - sampler uTextureSmplr [[id(2)]]; -}; - struct main0_out { float4 oFragColor [[color(0)]]; @@ -21,12 +14,12 @@ struct main0_in float2 vTexCoord [[user(locn0)]]; }; -fragment main0_out main0(main0_in in [[stage_in]], constant spvDescriptorSetBuffer0& spvDescriptorSet0 [[buffer(0)]]) +fragment main0_out main0(main0_in in [[stage_in]], constant float4x4& uOldTransform [[buffer(0)]], texture2d uTexture [[texture(0)]], sampler uTextureSmplr [[sampler(0)]]) { main0_out out = {}; - float4 normTexCoord = (*spvDescriptorSet0.uOldTransform) * float4(in.vTexCoord, 0.0, 1.0); + float4 normTexCoord = uOldTransform * float4(in.vTexCoord, 0.0, 1.0); float2 texCoord = ((normTexCoord.xy / float2(normTexCoord.w)) + float2(1.0)) * 0.5; - out.oFragColor = spvDescriptorSet0.uTexture.sample(spvDescriptorSet0.uTextureSmplr, texCoord); + out.oFragColor = uTexture.sample(uTextureSmplr, texCoord); return out; } diff --git a/resources/shaders/metal/reproject.vs.metal b/resources/shaders/metal/reproject.vs.metal index 3bdc9034..da03751f 100644 --- a/resources/shaders/metal/reproject.vs.metal +++ b/resources/shaders/metal/reproject.vs.metal @@ -4,11 +4,6 @@ using namespace metal; -struct spvDescriptorSetBuffer0 -{ - constant float4x4* uNewTransform [[id(0)]]; -}; - struct main0_out { float2 vTexCoord [[user(locn0)]]; @@ -20,13 +15,13 @@ struct main0_in int2 aPosition [[attribute(0)]]; }; -vertex main0_out main0(main0_in in [[stage_in]], constant spvDescriptorSetBuffer0& spvDescriptorSet0 [[buffer(0)]]) +vertex main0_out main0(main0_in in [[stage_in]], constant float4x4& uNewTransform [[buffer(0)]]) { main0_out out = {}; float2 position = float2(in.aPosition); out.vTexCoord = position; position.y = 1.0 - position.y; - out.gl_Position = (*spvDescriptorSet0.uNewTransform) * float4(position, 0.0, 1.0); + out.gl_Position = uNewTransform * float4(position, 0.0, 1.0); return out; } diff --git a/resources/shaders/metal/tile.fs.metal b/resources/shaders/metal/tile.fs.metal index 317a2cb7..fcb15dd7 100644 --- a/resources/shaders/metal/tile.fs.metal +++ b/resources/shaders/metal/tile.fs.metal @@ -6,24 +6,6 @@ using namespace metal; -struct spvDescriptorSetBuffer0 -{ - texture2d uMaskTexture0 [[id(0)]]; - sampler uMaskTexture0Smplr [[id(1)]]; - texture2d uColorTexture0 [[id(2)]]; - sampler uColorTexture0Smplr [[id(3)]]; - texture2d uGammaLUT [[id(4)]]; - sampler uGammaLUTSmplr [[id(5)]]; - constant float2* uColorTexture0Size [[id(6)]]; - constant float2* uFramebufferSize [[id(7)]]; - constant float4* uFilterParams0 [[id(8)]]; - constant float4* uFilterParams1 [[id(9)]]; - constant float4* uFilterParams2 [[id(10)]]; - texture2d uDestTexture [[id(11)]]; - sampler uDestTextureSmplr [[id(12)]]; - constant int* uCtrl [[id(13)]]; -}; - constant float3 _1041 = {}; struct main0_out @@ -41,11 +23,12 @@ struct main0_in // Implementation of the GLSL mod() function, which is slightly different than Metal fmod() template -Tx mod(Tx x, Ty y) +inline Tx mod(Tx x, Ty y) { return x - y * floor(x / y); } +static inline __attribute__((always_inline)) float sampleMask(thread const float& maskAlpha, thread const texture2d maskTexture, thread const sampler maskTextureSmplr, thread const float3& maskTexCoord, thread const int& maskCtrl) { if (maskCtrl == 0) @@ -64,6 +47,7 @@ float sampleMask(thread const float& maskAlpha, thread const texture2d ma return fast::min(maskAlpha, coverage); } +static inline __attribute__((always_inline)) float4 filterRadialGradient(thread const float2& colorTexCoord, thread const texture2d colorTexture, thread const sampler colorTextureSmplr, thread const float2& colorTextureSize, thread const float2& fragCoord, thread const float2& framebufferSize, thread const float4& filterParams0, thread const float4& filterParams1) { float2 lineFrom = filterParams0.xy; @@ -100,6 +84,7 @@ float4 filterRadialGradient(thread const float2& colorTexCoord, thread const tex return color; } +static inline __attribute__((always_inline)) float4 filterBlur(thread const float2& colorTexCoord, thread const texture2d colorTexture, thread const sampler colorTextureSmplr, thread const float2& colorTextureSize, thread const float4& filterParams0, thread const float4& filterParams1) { float2 srcOffsetScale = filterParams0.xy / colorTextureSize; @@ -124,11 +109,13 @@ float4 filterBlur(thread const float2& colorTexCoord, thread const texture2d colorTexture, thread const sampler colorTextureSmplr, thread const float2& colorTexCoord) { return colorTexture.sample(colorTextureSmplr, (colorTexCoord + float2(offset, 0.0))).x; } +static inline __attribute__((always_inline)) void filterTextSample9Tap(thread float4& outAlphaLeft, thread float& outAlphaCenter, thread float4& outAlphaRight, thread const texture2d colorTexture, thread const sampler colorTextureSmplr, thread const float2& colorTexCoord, thread const float4& kernel0, thread const float& onePixel) { bool wide = kernel0.x > 0.0; @@ -173,16 +160,19 @@ void filterTextSample9Tap(thread float4& outAlphaLeft, thread float& outAlphaCen outAlphaRight = float4(filterTextSample1Tap(param_10, colorTexture, colorTextureSmplr, param_11), filterTextSample1Tap(param_12, colorTexture, colorTextureSmplr, param_13), filterTextSample1Tap(param_14, colorTexture, colorTextureSmplr, param_15), _295); } +static inline __attribute__((always_inline)) float filterTextConvolve7Tap(thread const float4& alpha0, thread const float3& alpha1, thread const float4& kernel0) { return dot(alpha0, kernel0) + dot(alpha1, kernel0.zyx); } +static inline __attribute__((always_inline)) float filterTextGammaCorrectChannel(thread const float& bgColor, thread const float& fgColor, thread const texture2d gammaLUT, thread const sampler gammaLUTSmplr) { return gammaLUT.sample(gammaLUTSmplr, float2(fgColor, 1.0 - bgColor)).x; } +static inline __attribute__((always_inline)) float3 filterTextGammaCorrect(thread const float3& bgColor, thread const float3& fgColor, thread const texture2d gammaLUT, thread const sampler gammaLUTSmplr) { float param = bgColor.x; @@ -194,6 +184,7 @@ float3 filterTextGammaCorrect(thread const float3& bgColor, thread const float3& return float3(filterTextGammaCorrectChannel(param, param_1, gammaLUT, gammaLUTSmplr), filterTextGammaCorrectChannel(param_2, param_3, gammaLUT, gammaLUTSmplr), filterTextGammaCorrectChannel(param_4, param_5, gammaLUT, gammaLUTSmplr)); } +static inline __attribute__((always_inline)) float4 filterText(thread const float2& colorTexCoord, thread const texture2d colorTexture, thread const sampler colorTextureSmplr, thread const texture2d gammaLUT, thread const sampler gammaLUTSmplr, thread const float2& colorTextureSize, thread const float4& filterParams0, thread const float4& filterParams1, thread const float4& filterParams2) { float4 kernel0 = filterParams0; @@ -240,17 +231,20 @@ float4 filterText(thread const float2& colorTexCoord, thread const texture2d colorTexture, thread const sampler colorTextureSmplr, thread const float2& colorTexCoord) { return colorTexture.sample(colorTextureSmplr, colorTexCoord); } +static inline __attribute__((always_inline)) float4 filterNone(thread const float2& colorTexCoord, thread const texture2d colorTexture, thread const sampler colorTextureSmplr) { float2 param = colorTexCoord; return sampleColor(colorTexture, colorTextureSmplr, param); } +static inline __attribute__((always_inline)) float4 filterColor(thread const float2& colorTexCoord, thread const texture2d colorTexture, thread const sampler colorTextureSmplr, thread const texture2d gammaLUT, thread const sampler gammaLUTSmplr, thread const float2& colorTextureSize, thread const float2& fragCoord, thread const float2& framebufferSize, thread const float4& filterParams0, thread const float4& filterParams1, thread const float4& filterParams2, thread const int& colorFilter) { switch (colorFilter) @@ -287,6 +281,7 @@ float4 filterColor(thread const float2& colorTexCoord, thread const texture2d destTexture, thread const sampler destTextureSmplr, thread const float2& destTextureSize, thread const float2& fragCoord, thread const int& op) { if (op == 0) @@ -546,6 +552,7 @@ float4 composite(thread const float4& srcColor, thread const texture2d de return float4(((srcColor.xyz * (srcColor.w * (1.0 - destColor.w))) + (blendedRGB * (srcColor.w * destColor.w))) + (destColor.xyz * (1.0 - srcColor.w)), 1.0); } +static inline __attribute__((always_inline)) void calculateColor(thread const int& tileCtrl, thread const int& ctrl, thread texture2d uMaskTexture0, thread const sampler uMaskTexture0Smplr, thread float3& vMaskTexCoord0, thread float4& vBaseColor, thread float2& vColorTexCoord0, thread texture2d uColorTexture0, thread const sampler uColorTexture0Smplr, thread texture2d uGammaLUT, thread const sampler uGammaLUTSmplr, thread float2 uColorTexture0Size, thread float4& gl_FragCoord, thread float2 uFramebufferSize, thread float4 uFilterParams0, thread float4 uFilterParams1, thread float4 uFilterParams2, thread texture2d uDestTexture, thread const sampler uDestTextureSmplr, thread float4& oFragColor) { int maskCtrl0 = (tileCtrl >> 0) & 3; @@ -585,12 +592,12 @@ void calculateColor(thread const int& tileCtrl, thread const int& ctrl, thread t oFragColor = color; } -fragment main0_out main0(main0_in in [[stage_in]], constant spvDescriptorSetBuffer0& spvDescriptorSet0 [[buffer(0)]], float4 gl_FragCoord [[position]]) +fragment main0_out main0(main0_in in [[stage_in]], constant int& uCtrl [[buffer(5)]], constant float2& uColorTexture0Size [[buffer(0)]], constant float2& uFramebufferSize [[buffer(1)]], constant float4& uFilterParams0 [[buffer(2)]], constant float4& uFilterParams1 [[buffer(3)]], constant float4& uFilterParams2 [[buffer(4)]], texture2d uMaskTexture0 [[texture(0)]], texture2d uColorTexture0 [[texture(1)]], texture2d uGammaLUT [[texture(2)]], texture2d uDestTexture [[texture(3)]], sampler uMaskTexture0Smplr [[sampler(0)]], sampler uColorTexture0Smplr [[sampler(1)]], sampler uGammaLUTSmplr [[sampler(2)]], sampler uDestTextureSmplr [[sampler(3)]], float4 gl_FragCoord [[position]]) { main0_out out = {}; int param = int(in.vTileCtrl); - int param_1 = (*spvDescriptorSet0.uCtrl); - calculateColor(param, param_1, spvDescriptorSet0.uMaskTexture0, spvDescriptorSet0.uMaskTexture0Smplr, in.vMaskTexCoord0, in.vBaseColor, in.vColorTexCoord0, spvDescriptorSet0.uColorTexture0, spvDescriptorSet0.uColorTexture0Smplr, spvDescriptorSet0.uGammaLUT, spvDescriptorSet0.uGammaLUTSmplr, (*spvDescriptorSet0.uColorTexture0Size), gl_FragCoord, (*spvDescriptorSet0.uFramebufferSize), (*spvDescriptorSet0.uFilterParams0), (*spvDescriptorSet0.uFilterParams1), (*spvDescriptorSet0.uFilterParams2), spvDescriptorSet0.uDestTexture, spvDescriptorSet0.uDestTextureSmplr, out.oFragColor); + int param_1 = uCtrl; + calculateColor(param, param_1, uMaskTexture0, uMaskTexture0Smplr, in.vMaskTexCoord0, in.vBaseColor, in.vColorTexCoord0, uColorTexture0, uColorTexture0Smplr, uGammaLUT, uGammaLUTSmplr, uColorTexture0Size, gl_FragCoord, uFramebufferSize, uFilterParams0, uFilterParams1, uFilterParams2, uDestTexture, uDestTextureSmplr, out.oFragColor); return out; } diff --git a/resources/shaders/metal/tile.vs.metal b/resources/shaders/metal/tile.vs.metal index 80e7ff82..f05bf6ae 100644 --- a/resources/shaders/metal/tile.vs.metal +++ b/resources/shaders/metal/tile.vs.metal @@ -4,15 +4,6 @@ using namespace metal; -struct spvDescriptorSetBuffer0 -{ - constant float2* uTileSize [[id(0)]]; - constant int2* uTextureMetadataSize [[id(1)]]; - texture2d uTextureMetadata [[id(2)]]; - sampler uTextureMetadataSmplr [[id(3)]]; - constant float4x4* uTransform [[id(4)]]; -}; - struct main0_out { float3 vMaskTexCoord0 [[user(locn0)]]; @@ -32,26 +23,26 @@ struct main0_in int aTileCtrl [[attribute(5)]]; }; -vertex main0_out main0(main0_in in [[stage_in]], constant spvDescriptorSetBuffer0& spvDescriptorSet0 [[buffer(0)]]) +vertex main0_out main0(main0_in in [[stage_in]], constant int2& uTextureMetadataSize [[buffer(1)]], constant float2& uTileSize [[buffer(0)]], constant float4x4& uTransform [[buffer(2)]], texture2d uTextureMetadata [[texture(0)]], sampler uTextureMetadataSmplr [[sampler(0)]]) { main0_out out = {}; float2 tileOrigin = float2(in.aTileOrigin); float2 tileOffset = float2(in.aTileOffset); - float2 position = (tileOrigin + tileOffset) * (*spvDescriptorSet0.uTileSize); + float2 position = (tileOrigin + tileOffset) * uTileSize; float2 maskTexCoord0 = (float2(in.aMaskTexCoord0) + tileOffset) / float2(256.0); - float2 textureMetadataScale = float2(1.0) / float2((*spvDescriptorSet0.uTextureMetadataSize)); + float2 textureMetadataScale = float2(1.0) / float2(uTextureMetadataSize); float2 metadataEntryCoord = float2(float((in.aColor % 128) * 4), float(in.aColor / 128)); float2 colorTexMatrix0Coord = (metadataEntryCoord + float2(0.5)) * textureMetadataScale; float2 colorTexOffsetsCoord = (metadataEntryCoord + float2(1.5, 0.5)) * textureMetadataScale; float2 baseColorCoord = (metadataEntryCoord + float2(2.5, 0.5)) * textureMetadataScale; - float4 colorTexMatrix0 = spvDescriptorSet0.uTextureMetadata.sample(spvDescriptorSet0.uTextureMetadataSmplr, colorTexMatrix0Coord, level(0.0)); - float4 colorTexOffsets = spvDescriptorSet0.uTextureMetadata.sample(spvDescriptorSet0.uTextureMetadataSmplr, colorTexOffsetsCoord, level(0.0)); - float4 baseColor = spvDescriptorSet0.uTextureMetadata.sample(spvDescriptorSet0.uTextureMetadataSmplr, baseColorCoord, level(0.0)); + float4 colorTexMatrix0 = uTextureMetadata.sample(uTextureMetadataSmplr, colorTexMatrix0Coord, level(0.0)); + float4 colorTexOffsets = uTextureMetadata.sample(uTextureMetadataSmplr, colorTexOffsetsCoord, level(0.0)); + float4 baseColor = uTextureMetadata.sample(uTextureMetadataSmplr, baseColorCoord, level(0.0)); out.vColorTexCoord0 = (float2x2(float2(colorTexMatrix0.xy), float2(colorTexMatrix0.zw)) * position) + colorTexOffsets.xy; out.vMaskTexCoord0 = float3(maskTexCoord0, float(in.aMaskBackdrop.x)); out.vBaseColor = baseColor; out.vTileCtrl = float(in.aTileCtrl); - out.gl_Position = (*spvDescriptorSet0.uTransform) * float4(position, 0.0, 1.0); + out.gl_Position = uTransform * float4(position, 0.0, 1.0); return out; } diff --git a/resources/shaders/metal/tile_clip.fs.metal b/resources/shaders/metal/tile_clip.fs.metal index a77450f0..f35707f0 100644 --- a/resources/shaders/metal/tile_clip.fs.metal +++ b/resources/shaders/metal/tile_clip.fs.metal @@ -4,12 +4,6 @@ using namespace metal; -struct spvDescriptorSetBuffer0 -{ - texture2d uSrc [[id(0)]]; - sampler uSrcSmplr [[id(1)]]; -}; - struct main0_out { float4 oFragColor [[color(0)]]; @@ -21,10 +15,10 @@ struct main0_in float vBackdrop [[user(locn1)]]; }; -fragment main0_out main0(main0_in in [[stage_in]], constant spvDescriptorSetBuffer0& spvDescriptorSet0 [[buffer(0)]]) +fragment main0_out main0(main0_in in [[stage_in]], texture2d uSrc [[texture(0)]], sampler uSrcSmplr [[sampler(0)]]) { main0_out out = {}; - float alpha = fast::clamp(abs(spvDescriptorSet0.uSrc.sample(spvDescriptorSet0.uSrcSmplr, in.vTexCoord).x + in.vBackdrop), 0.0, 1.0); + float alpha = fast::clamp(abs(uSrc.sample(uSrcSmplr, in.vTexCoord).x + in.vBackdrop), 0.0, 1.0); out.oFragColor = float4(alpha, 0.0, 0.0, 1.0); return out; } diff --git a/resources/shaders/metal/tile_copy.fs.metal b/resources/shaders/metal/tile_copy.fs.metal index fac13539..67e61bef 100644 --- a/resources/shaders/metal/tile_copy.fs.metal +++ b/resources/shaders/metal/tile_copy.fs.metal @@ -4,23 +4,16 @@ using namespace metal; -struct spvDescriptorSetBuffer0 -{ - constant float2* uFramebufferSize [[id(0)]]; - texture2d uSrc [[id(1)]]; - sampler uSrcSmplr [[id(2)]]; -}; - struct main0_out { float4 oFragColor [[color(0)]]; }; -fragment main0_out main0(constant spvDescriptorSetBuffer0& spvDescriptorSet0 [[buffer(0)]], float4 gl_FragCoord [[position]]) +fragment main0_out main0(constant float2& uFramebufferSize [[buffer(0)]], texture2d uSrc [[texture(0)]], sampler uSrcSmplr [[sampler(0)]], float4 gl_FragCoord [[position]]) { main0_out out = {}; - float2 texCoord = gl_FragCoord.xy / (*spvDescriptorSet0.uFramebufferSize); - out.oFragColor = spvDescriptorSet0.uSrc.sample(spvDescriptorSet0.uSrcSmplr, texCoord); + float2 texCoord = gl_FragCoord.xy / uFramebufferSize; + out.oFragColor = uSrc.sample(uSrcSmplr, texCoord); return out; } diff --git a/resources/shaders/metal/tile_copy.vs.metal b/resources/shaders/metal/tile_copy.vs.metal index 530c42d7..6ac89c86 100644 --- a/resources/shaders/metal/tile_copy.vs.metal +++ b/resources/shaders/metal/tile_copy.vs.metal @@ -4,12 +4,6 @@ using namespace metal; -struct spvDescriptorSetBuffer0 -{ - constant float2* uTileSize [[id(0)]]; - constant float4x4* uTransform [[id(1)]]; -}; - struct main0_out { float4 gl_Position [[position]]; @@ -20,11 +14,11 @@ struct main0_in int2 aTilePosition [[attribute(0)]]; }; -vertex main0_out main0(main0_in in [[stage_in]], constant spvDescriptorSetBuffer0& spvDescriptorSet0 [[buffer(0)]]) +vertex main0_out main0(main0_in in [[stage_in]], constant float2& uTileSize [[buffer(0)]], constant float4x4& uTransform [[buffer(1)]]) { main0_out out = {}; - float2 position = float2(in.aTilePosition) * (*spvDescriptorSet0.uTileSize); - out.gl_Position = (*spvDescriptorSet0.uTransform) * float4(position, 0.0, 1.0); + float2 position = float2(in.aTilePosition) * uTileSize; + out.gl_Position = uTransform * float4(position, 0.0, 1.0); return out; } diff --git a/shaders/Makefile b/shaders/Makefile index f42582ad..9002a03b 100644 --- a/shaders/Makefile +++ b/shaders/Makefile @@ -25,21 +25,29 @@ SHADERS=\ tile_copy.vs.glsl \ $(EMPTY) +COMPUTE_SHADERS=\ + fill.cs.glsl \ + $(EMPTY) + INCLUDES=\ + fill.inc.glsl \ $(EMPTY) OUT=\ $(SHADERS:%=$(TARGET_DIR)/gl3/%) \ $(SHADERS:%.glsl=$(TARGET_DIR)/metal/%.metal) \ $(SHADERS:%.glsl=build/metal/%.spv) \ + $(COMPUTE_SHADERS:%.glsl=$(TARGET_DIR)/metal/%.metal) \ + $(COMPUTE_SHADERS:%.glsl=build/metal/%.spv) \ $(EMPTY) GLSL_VERSION=330 +GLSL_COMPUTE_VERSION=430 GLSLANGFLAGS=--auto-map-locations -I. GLSLANGFLAGS_METAL=$(GLSLANGFLAGS) -DPF_ORIGIN_UPPER_LEFT=1 SPIRVCROSS?=spirv-cross -SPIRVCROSSFLAGS=--msl --msl-version 020100 --msl-argument-buffers +SPIRVCROSSFLAGS=--msl --msl-version 020100 GLSL_VERSION_HEADER="\#version {{version}}" HEADER="// Automatically generated from files in pathfinder/shaders/. Do not edit!" @@ -65,5 +73,8 @@ build/metal/%.vs.spv: %.vs.glsl $(INCLUDES) $(TARGET_DIR)/gl3/%.vs.glsl: %.vs.glsl $(INCLUDES) mkdir -p $(TARGET_DIR)/gl3 && echo $(GLSL_VERSION_HEADER) > $@ && echo $(HEADER) >> $@ && ( glslangValidator $(GLSLANGFLAGS) -S vert -E $< | sed $(GLSL_SED_ARGS) >> $@ ) || ( rm $@ && exit 1 ) +build/metal/%.cs.spv: %.cs.glsl $(INCLUDES) + mkdir -p build/metal && glslangValidator $(GLSLANGFLAGS_METAL) -G$(GLSL_COMPUTE_VERSION) -S comp -o $@ $< + $(TARGET_DIR)/metal/%.metal: build/metal/%.spv mkdir -p $(TARGET_DIR)/metal && echo $(HEADER) > $@ && ( $(SPIRVCROSS) $(SPIRVCROSSFLAGS) $< | sed $(METAL_SED_ARGS) >> $@ ) || ( rm $@ && exit 1 ) diff --git a/shaders/fill.cs.glsl b/shaders/fill.cs.glsl new file mode 100644 index 00000000..93bdcecb --- /dev/null +++ b/shaders/fill.cs.glsl @@ -0,0 +1,67 @@ +#version 430 + +// pathfinder/shaders/fill.cs.glsl +// +// Copyright © 2020 The Pathfinder Project Developers. +// +// Licensed under the Apache License, Version 2.0 or the MIT license +// , at your +// option. This file may not be copied, modified, or distributed +// except according to those terms. + +#extension GL_GOOGLE_include_directive : enable + +precision highp float; +precision highp sampler2D; + +#include "fill.inc.glsl" + +layout(local_size_x = 16, local_size_y = 16) in; + +uniform writeonly image2D uDest; +uniform sampler2D uAreaLUT; +uniform int uFirstTileIndex; + +layout(std430, binding = 0) buffer bFills { + restrict readonly uvec2 iFills[]; +}; + +layout(std430, binding = 1) buffer bNextFills { + restrict readonly int iNextFills[]; +}; + +layout(std430, binding = 2) buffer bFillTileMap { + restrict readonly int iFillTileMap[]; +}; + +void main() { + ivec2 tileSubCoord = ivec2(gl_LocalInvocationID.xy); + uint tileIndexOffset = gl_WorkGroupID.z; + + uint tileIndex = tileIndexOffset + uint(uFirstTileIndex); + ivec2 tileOrigin = ivec2(tileIndex & 0xff, (tileIndex >> 8u) & 0xff) * 16; + ivec2 destCoord = tileOrigin + tileSubCoord; + + int fillIndex = iFillTileMap[tileIndex]; + if (fillIndex < 0) + return; + + float coverage = 0.0; + do { + uvec2 fill = iFills[fillIndex]; + vec2 from = vec2(fill.y & 0xf, (fill.y >> 4u) & 0xf) + + vec2(fill.x & 0xff, (fill.x >> 8u) & 0xff) / 256.0; + vec2 to = vec2((fill.y >> 8u) & 0xf, (fill.y >> 12u) & 0xf) + + vec2((fill.x >> 16u) & 0xff, (fill.x >> 24u) & 0xff) / 256.0; + + from -= vec2(tileSubCoord) + vec2(0.5); + to -= vec2(tileSubCoord) + vec2(0.5); + + coverage += computeCoverage(from, to, uAreaLUT); + + fillIndex = iNextFills[fillIndex]; + } while (fillIndex >= 0); + + imageStore(uDest, destCoord, vec4(coverage)); +} diff --git a/shaders/fill.fs.glsl b/shaders/fill.fs.glsl index 89dae088..89da3e01 100644 --- a/shaders/fill.fs.glsl +++ b/shaders/fill.fs.glsl @@ -1,8 +1,8 @@ #version 330 -// pathfinder/shaders/stencil.fs.glsl +// pathfinder/shaders/fill.fs.glsl // -// Copyright © 2019 The Pathfinder Project Developers. +// Copyright © 2020 The Pathfinder Project Developers. // // Licensed under the Apache License, Version 2.0 or the MIT license @@ -10,9 +10,13 @@ // option. This file may not be copied, modified, or distributed // except according to those terms. +#extension GL_GOOGLE_include_directive : enable + precision highp float; precision highp sampler2D; +#include "fill.inc.glsl" + uniform sampler2D uAreaLUT; in vec2 vFrom; @@ -21,22 +25,5 @@ in vec2 vTo; out vec4 oFragColor; void main() { - // Unpack. - vec2 from = vFrom, to = vTo; - - // Determine winding, and sort into a consistent order so we only need to find one root below. - vec2 left = from.x < to.x ? from : to, right = from.x < to.x ? to : from; - - // Shoot a vertical ray toward the curve. - vec2 window = clamp(vec2(from.x, to.x), -0.5, 0.5); - float offset = mix(window.x, window.y, 0.5) - left.x; - float t = offset / (right.x - left.x); - - // Compute position and derivative to form a line approximation. - float y = mix(left.y, right.y, t); - float d = (right.y - left.y) / (right.x - left.x); - - // Look up area under that line, and scale horizontally to the window size. - float dX = window.x - window.y; - oFragColor = vec4(texture(uAreaLUT, vec2(y + 8.0, abs(d * dX)) / 16.0).r * dX); + oFragColor = vec4(computeCoverage(vFrom, vTo, uAreaLUT)); } diff --git a/shaders/fill.inc.glsl b/shaders/fill.inc.glsl new file mode 100644 index 00000000..356248f1 --- /dev/null +++ b/shaders/fill.inc.glsl @@ -0,0 +1,27 @@ +// pathfinder/shaders/fill.inc.glsl +// +// Copyright © 2020 The Pathfinder Project Developers. +// +// Licensed under the Apache License, Version 2.0 or the MIT license +// , at your +// option. This file may not be copied, modified, or distributed +// except according to those terms. + +float computeCoverage(vec2 from, vec2 to, sampler2D areaLUT) { + // Determine winding, and sort into a consistent order so we only need to find one root below. + vec2 left = from.x < to.x ? from : to, right = from.x < to.x ? to : from; + + // Shoot a vertical ray toward the curve. + vec2 window = clamp(vec2(from.x, to.x), -0.5, 0.5); + float offset = mix(window.x, window.y, 0.5) - left.x; + float t = offset / (right.x - left.x); + + // Compute position and derivative to form a line approximation. + float y = mix(left.y, right.y, t); + float d = (right.y - left.y) / (right.x - left.x); + + // Look up area under that line, and scale horizontally to the window size. + float dX = window.x - window.y; + return texture(areaLUT, vec2(y + 8.0, abs(d * dX)) / 16.0).r * dX; +} diff --git a/shaders/fill.vs.glsl b/shaders/fill.vs.glsl index d82e8688..20d04463 100644 --- a/shaders/fill.vs.glsl +++ b/shaders/fill.vs.glsl @@ -2,7 +2,7 @@ // pathfinder/shaders/fill.vs.glsl // -// Copyright © 2019 The Pathfinder Project Developers. +// Copyright © 2020 The Pathfinder Project Developers. // // Licensed under the Apache License, Version 2.0 or the MIT license diff --git a/ui/src/lib.rs b/ui/src/lib.rs index 74803af0..35322543 100644 --- a/ui/src/lib.rs +++ b/ui/src/lib.rs @@ -166,12 +166,10 @@ impl UIPresenter where D: Device { filled: bool) { device.allocate_buffer(&self.solid_vertex_array.vertex_buffer, BufferData::Memory(vertex_data), - BufferTarget::Vertex, - BufferUploadMode::Dynamic); + BufferTarget::Vertex); device.allocate_buffer(&self.solid_vertex_array.index_buffer, BufferData::Memory(index_data), - BufferTarget::Index, - BufferUploadMode::Dynamic); + BufferTarget::Index); let primitive = if filled { Primitive::Triangles } else { Primitive::Lines }; device.draw_elements(index_data.len() as u32, &RenderState { @@ -185,6 +183,7 @@ impl UIPresenter where D: Device { (&self.solid_program.color_uniform, get_color_uniform(color)), ], textures: &[], + images: &[], viewport: RectI::new(Vector2I::default(), self.framebuffer_size), options: RenderOptions { blend: Some(alpha_blend_state()), @@ -398,12 +397,10 @@ impl UIPresenter where D: Device { color: ColorU) { device.allocate_buffer(&self.texture_vertex_array.vertex_buffer, BufferData::Memory(vertex_data), - BufferTarget::Vertex, - BufferUploadMode::Dynamic); + BufferTarget::Vertex); device.allocate_buffer(&self.texture_vertex_array.index_buffer, BufferData::Memory(index_data), - BufferTarget::Index, - BufferUploadMode::Dynamic); + BufferTarget::Index); device.draw_elements(index_data.len() as u32, &RenderState { target: &RenderTarget::Default, @@ -411,6 +408,7 @@ impl UIPresenter where D: Device { vertex_array: &self.texture_vertex_array.vertex_array, primitive: Primitive::Triangles, textures: &[&texture], + images: &[], uniforms: &[ (&self.texture_program.framebuffer_size_uniform, UniformData::Vec2(self.framebuffer_size.0.to_f32x2())), @@ -569,7 +567,7 @@ struct DebugTextureProgram where D: Device { impl DebugTextureProgram where D: Device { fn new(device: &D, resources: &dyn ResourceLoader) -> DebugTextureProgram { - let program = device.create_program(resources, "debug_texture"); + let program = device.create_raster_program(resources, "debug_texture"); let framebuffer_size_uniform = device.get_uniform(&program, "FramebufferSize"); let texture_size_uniform = device.get_uniform(&program, "TextureSize"); let texture_uniform = device.get_uniform(&program, "Texture"); @@ -593,7 +591,8 @@ struct DebugTextureVertexArray where D: Device { impl DebugTextureVertexArray where D: Device { fn new(device: &D, debug_texture_program: &DebugTextureProgram) -> DebugTextureVertexArray { - let (vertex_buffer, index_buffer) = (device.create_buffer(), device.create_buffer()); + let vertex_buffer = device.create_buffer(BufferUploadMode::Dynamic); + let index_buffer = device.create_buffer(BufferUploadMode::Dynamic); let vertex_array = device.create_vertex_array(); let position_attr = device.get_vertex_attr(&debug_texture_program.program, "Position") @@ -634,7 +633,8 @@ struct DebugSolidVertexArray where D: Device { impl DebugSolidVertexArray where D: Device { fn new(device: &D, debug_solid_program: &DebugSolidProgram) -> DebugSolidVertexArray { - let (vertex_buffer, index_buffer) = (device.create_buffer(), device.create_buffer()); + let vertex_buffer = device.create_buffer(BufferUploadMode::Dynamic); + let index_buffer = device.create_buffer(BufferUploadMode::Dynamic); let vertex_array = device.create_vertex_array(); let position_attr = @@ -663,7 +663,7 @@ struct DebugSolidProgram where D: Device { impl DebugSolidProgram where D: Device { fn new(device: &D, resources: &dyn ResourceLoader) -> DebugSolidProgram { - let program = device.create_program(resources, "debug_solid"); + let program = device.create_raster_program(resources, "debug_solid"); let framebuffer_size_uniform = device.get_uniform(&program, "FramebufferSize"); let color_uniform = device.get_uniform(&program, "Color"); DebugSolidProgram { program, framebuffer_size_uniform, color_uniform } diff --git a/webgl/src/lib.rs b/webgl/src/lib.rs index 42b8291b..37c1e9dc 100644 --- a/webgl/src/lib.rs +++ b/webgl/src/lib.rs @@ -16,15 +16,15 @@ extern crate log; use pathfinder_geometry::rect::RectI; use pathfinder_geometry::vector::Vector2I; use pathfinder_gpu::{BlendFactor, BlendOp, BufferData, BufferTarget, RenderTarget}; -use pathfinder_gpu::{BufferUploadMode, ClearOps, DepthFunc, Device, Primitive, RenderOptions}; -use pathfinder_gpu::{RenderState, ShaderKind, StencilFunc, TextureData, TextureDataRef}; -use pathfinder_gpu::{TextureFormat, TextureSamplingFlags, UniformData, VertexAttrClass}; -use pathfinder_gpu::{VertexAttrDescriptor, VertexAttrType}; +use pathfinder_gpu::{BufferUploadMode, ClearOps, ComputeDimensions, ComputeState, DepthFunc, Device, Primitive, ProgramKind}; +use pathfinder_gpu::{RenderOptions, RenderState, ShaderKind, StencilFunc, TextureData}; +use pathfinder_gpu::{TextureDataRef, TextureFormat, TextureSamplingFlags, UniformData}; +use pathfinder_gpu::{VertexAttrClass, VertexAttrDescriptor, VertexAttrType}; use pathfinder_resources::ResourceLoader; use std::mem; use std::str; use std::time::Duration; -use wasm_bindgen::{JsCast, JsValue}; +use wasm_bindgen::JsCast; use web_sys::WebGl2RenderingContext as WebGl; use js_sys::{Uint8Array, Uint16Array, Float32Array, Object}; @@ -169,7 +169,7 @@ impl WebGlDevice { self.context.uniform3i(location, data[0], data[1], data[2]); self.ck(); } - UniformData::TextureUnit(unit) => { + UniformData::TextureUnit(unit) | UniformData::ImageUnit(unit) => { self.context.uniform1i(location, unit as i32); self.ck(); } @@ -412,9 +412,11 @@ unsafe fn check_and_extract_data( impl Device for WebGlDevice { type Buffer = WebGlBuffer; + type Fence = (); type Framebuffer = WebGlFramebuffer; type Program = WebGlProgram; type Shader = WebGlShader; + type StorageBuffer = (); type Texture = WebGlTexture; type TextureDataReceiver = (); type TimerQuery = WebGlTimerQuery; @@ -504,6 +506,7 @@ impl Device for WebGlDevice { let gl_shader_kind = match kind { ShaderKind::Vertex => WebGl::VERTEX_SHADER, ShaderKind::Fragment => WebGl::FRAGMENT_SHADER, + ShaderKind::Compute => panic!("Compute shaders are unsupported in WebGL!"), }; let gl_shader = self @@ -529,17 +532,21 @@ impl Device for WebGlDevice { &self, _resources: &dyn ResourceLoader, name: &str, - vertex_shader: WebGlShader, - fragment_shader: WebGlShader, + shaders: ProgramKind, ) -> WebGlProgram { let gl_program = self .context .create_program() .expect("unable to create program object"); - self.context - .attach_shader(&gl_program, &vertex_shader.gl_shader); - self.context - .attach_shader(&gl_program, &fragment_shader.gl_shader); + match shaders { + ProgramKind::Raster { ref vertex, ref fragment } => { + self.context.attach_shader(&gl_program, &vertex.gl_shader); + self.context.attach_shader(&gl_program, &fragment.gl_shader); + } + ProgramKind::Compute(ref shader) => { + self.context.attach_shader(&gl_program, &shader.gl_shader); + } + } self.context.link_program(&gl_program); if !self .context @@ -559,6 +566,11 @@ impl Device for WebGlDevice { } } + #[inline] + fn set_compute_program_local_size(&self, _: &mut Self::Program, _: ComputeDimensions) { + // This does nothing on OpenGL, since the local size is set in the shader. + } + #[inline] fn create_vertex_array(&self) -> WebGlVertexArray { WebGlVertexArray { @@ -585,6 +597,10 @@ impl Device for WebGlDevice { WebGlUniform { location: location } } + fn get_storage_buffer(&self, _: &Self::Program, _: &str, _: u32) { + // TODO(pcwalton) + } + fn configure_vertex_attr( &self, vertex_array: &WebGlVertexArray, @@ -671,11 +687,12 @@ impl Device for WebGlDevice { framebuffer.texture } - fn create_buffer(&self) -> WebGlBuffer { + fn create_buffer(&self, mode: BufferUploadMode) -> WebGlBuffer { let buffer = self.context.create_buffer().unwrap(); WebGlBuffer { buffer, context: self.context.clone(), + mode, } } @@ -684,15 +701,15 @@ impl Device for WebGlDevice { buffer: &WebGlBuffer, data: BufferData, target: BufferTarget, - mode: BufferUploadMode, ) { let target = match target { BufferTarget::Vertex => WebGl::ARRAY_BUFFER, BufferTarget::Index => WebGl::ELEMENT_ARRAY_BUFFER, + BufferTarget::Storage => panic!("Shader storage buffers are unsupported in WebGL!"), }; self.context.bind_buffer(target, Some(&buffer.buffer)); self.ck(); - let usage = mode.to_gl_usage(); + let usage = buffer.mode.to_gl_usage(); match data { BufferData::Uninitialized(len) => { self.context @@ -705,6 +722,18 @@ impl Device for WebGlDevice { } } + fn upload_to_buffer(&self, + buffer: &Self::Buffer, + position: usize, + data: &[T], + target: BufferTarget) { + let target = target.to_gl_target(); + self.context.bind_buffer(target, Some(&buffer.buffer)); self.ck(); + self.context.buffer_sub_data_with_i32_and_u8_array(target, + position as i32, + slice_to_u8(data)); self.ck(); + } + #[inline] fn framebuffer_texture<'f>(&self, framebuffer: &'f Self::Framebuffer) -> &'f Self::Texture { &framebuffer.texture @@ -845,6 +874,10 @@ impl Device for WebGlDevice { self.reset_render_state(render_state); } + fn dispatch_compute(&self, _: ComputeDimensions, _: &ComputeState) { + panic!("Compute shader is unsupported in WebGL!") + } + #[inline] fn create_timer_query(&self) -> WebGlTimerQuery { // FIXME use performance timers @@ -900,10 +933,19 @@ impl Device for WebGlDevice { let suffix = match kind { ShaderKind::Vertex => 'v', ShaderKind::Fragment => 'f', + ShaderKind::Compute => 'c', }; let path = format!("shaders/gl3/{}.{}s.glsl", name, suffix); self.create_shader_from_source(name, &resources.slurp(&path).unwrap(), kind) } + + fn add_fence(&self) -> Self::Fence { + // TODO(pcwalton) + } + + fn wait_for_fence(&self, _: &Self::Fence) { + // TODO(pcwalton) + } } pub struct WebGlVertexArray { @@ -931,6 +973,7 @@ pub struct WebGlFramebuffer { pub struct WebGlBuffer { context: web_sys::WebGl2RenderingContext, pub buffer: web_sys::WebGlBuffer, + pub mode: BufferUploadMode, } impl Drop for WebGlBuffer { @@ -982,6 +1025,7 @@ impl BufferTargetExt for BufferTarget { match self { BufferTarget::Vertex => WebGl::ARRAY_BUFFER, BufferTarget::Index => WebGl::ELEMENT_ARRAY_BUFFER, + BufferTarget::Storage => panic!("Shader storage buffers are unsupported in WebGL!"), } } }