feat: bring back WebGPU (#20812)

Signed-off-by: Leo Kettmeir <crowlkats@toaxl.com>
Co-authored-by: Kenta Moriuchi <moriken@kimamass.com>
Co-authored-by: Bartek Iwańczuk <biwanczuk@gmail.com>

.cargo/config.toml

@@ -11,8 +11,17 @@ rustflags = [
   "link-arg=/STACK:4194304",
 ]
 
+[target.x86_64-apple-darwin]
+rustflags = [
+  "-C",
+  "link-args=-weak_framework Metal -weak_framework MetalPerformanceShaders -weak_framework QuartzCore -weak_framework CoreGraphics",
+]
+
 [target.aarch64-apple-darwin]
-rustflags = ["-C", "link-arg=-fuse-ld=lld"]
+rustflags = [
+  "-C",
+  "link-args=-fuse-ld=lld -weak_framework Metal -weak_framework MetalPerformanceShaders -weak_framework QuartzCore -weak_framework CoreGraphics",
+]
 
 [target.'cfg(all())']
 rustflags = [

Cargo.toml

@@ -25,6 +25,7 @@ members = [
   "ext/node",
   "ext/url",
   "ext/web",
+  "ext/webgpu",
   "ext/webidl",
   "ext/websocket",
   "ext/webstorage",
@@ -71,6 +72,7 @@ deno_kv = { version = "0.35.0", path = "./ext/kv" }
 deno_tls = { version = "0.114.0", path = "./ext/tls" }
 deno_url = { version = "0.127.0", path = "./ext/url" }
 deno_web = { version = "0.158.0", path = "./ext/web" }
+deno_webgpu = { version = "0.94.0", path = "./ext/webgpu" }
 deno_webidl = { version = "0.127.0", path = "./ext/webidl" }
 deno_websocket = { version = "0.132.0", path = "./ext/websocket" }
 deno_webstorage = { version = "0.122.0", path = "./ext/webstorage" }
@@ -164,6 +166,12 @@ p384 = { version = "0.13.0", features = ["ecdh"] }
 rsa = { version = "0.9.3", default-features = false, features = ["std", "pem", "hazmat"] } # hazmat needed for PrehashSigner in ext/node
 hkdf = "0.12.3"
 
+# webgpu
+raw-window-handle = "0.5.0"
+wgpu-core = "=0.18"
+wgpu-types = "=0.18"
+wgpu-hal = "=0.18"
+
 # macros
 proc-macro2 = "1"
 quote = "1"

cli/args/flags.rs

@@ -815,6 +815,7 @@ static ENV_VARIABLES_HELP: &str = r#"ENVIRONMENT VARIABLES:
     DENO_NO_UPDATE_CHECK Set to disable checking if a newer Deno version is
                          available
     DENO_V8_FLAGS        Set V8 command line options
+    DENO_WEBGPU_TRACE    Directory to use for wgpu traces
     DENO_JOBS            Number of parallel workers used for the --parallel
                          flag with the test subcommand. Defaults to number
                          of available CPUs.

cli/build.rs

@@ -149,6 +149,7 @@ mod ts {
     op_crate_libs.insert("deno.url", deno_url::get_declaration());
     op_crate_libs.insert("deno.web", deno_web::get_declaration());
     op_crate_libs.insert("deno.fetch", deno_fetch::get_declaration());
+    op_crate_libs.insert("deno.webgpu", deno_webgpu_get_declaration());
     op_crate_libs.insert("deno.websocket", deno_websocket::get_declaration());
     op_crate_libs.insert("deno.webstorage", deno_webstorage::get_declaration());
     op_crate_libs.insert("deno.crypto", deno_crypto::get_declaration());
@@ -458,3 +459,11 @@ fn main() {
     res.compile().unwrap();
   }
 }
+
+fn deno_webgpu_get_declaration() -> PathBuf {
+  let manifest_dir = std::path::Path::new(env!("CARGO_MANIFEST_DIR"));
+  manifest_dir
+    .join("tsc")
+    .join("dts")
+    .join("lib.deno_webgpu.d.ts")
+}

cli/js/40_testing.js

@@ -142,7 +142,10 @@ const OP_DETAILS = {
     "op_utime_async": ["change file timestamps", "awaiting the result of a `Deno.utime` call"],
     "op_host_recv_message": ["receive a message from a web worker", "terminating a `Worker`"],
     "op_host_recv_ctrl": ["receive a message from a web worker", "terminating a `Worker`"],
-    "op_ws_close": ["close a WebSocket", "awaiting until the `close` event is emitted on a `WebSocket`, or the `WebSocketStream#closed` promise resolves"],
+    "op_webgpu_buffer_get_map_async": ["map a WebGPU buffer", "awaiting the result of a `GPUBuffer#mapAsync` call"],
+	  "op_webgpu_request_adapter": ["request a WebGPU adapter", "awaiting the result of a `navigator.gpu.requestAdapter` call"],
+	  "op_webgpu_request_device": ["request a WebGPU device", "awaiting the result of a `GPUAdapter#requestDevice` call"],
+	  "op_ws_close": ["close a WebSocket", "awaiting until the `close` event is emitted on a `WebSocket`, or the `WebSocketStream#closed` promise resolves"],
     "op_ws_create": ["create a WebSocket", "awaiting until the `open` event is emitted on a `WebSocket`, or the result of a `WebSocketStream#connection` promise"],
     "op_ws_next_event": ["receive the next message on a WebSocket", "closing a `WebSocket` or `WebSocketStream`"],
     "op_ws_send_text": ["send a message on a WebSocket", "closing a `WebSocket` or `WebSocketStream`"],

cli/main.rs

@@ -322,10 +322,15 @@ pub(crate) static UNSTABLE_GRANULAR_FLAGS: &[(
     // for "unstableFeatures" to see where it's used.
     8,
   ),
+  (
+    deno_runtime::deno_webgpu::UNSTABLE_FEATURE_NAME,
+    "Enable unstable `WebGPU` API",
+    9,
+  ),
   (
     deno_runtime::ops::worker_host::UNSTABLE_FEATURE_NAME,
     "Enable unstable Web Worker APIs",
-    9,
+    10,
   ),
 ];
 

cli/tests/integration/js_unit_tests.rs

@@ -101,6 +101,7 @@ util::unit_test_factory!(
     version_test,
     wasm_test,
     webcrypto_test,
+    webgpu_test,
     websocket_test,
     webstorage_test,
     worker_permissions_test,

cli/tests/integration/run_tests.rs

@@ -1660,6 +1660,17 @@ itest!(unstable_kv_enabled {
   output: "run/unstable_kv.enabled.out",
 });
 
+itest!(unstable_webgpu_disabled {
+  args: "run --quiet --reload --allow-read run/unstable_webgpu.js",
+  output: "run/unstable_webgpu.disabled.out",
+});
+
+itest!(unstable_webgpu_enabled {
+  args:
+    "run --quiet --reload --allow-read --unstable-webgpu run/unstable_webgpu.js",
+  output: "run/unstable_webgpu.enabled.out",
+});
+
 itest!(import_compression {
   args: "run --quiet --reload --allow-net run/import_compression/main.ts",
   output: "run/import_compression/main.out",

cli/tests/integration/shared_library_tests.rs

@@ -45,14 +45,25 @@ fn macos_shared_libraries() {
   // target/release/deno:
   //  /System/Library/Frameworks/CoreFoundation.framework/Versions/A/CoreFoundation (compatibility version 150.0.0, current version 1953.1.0)
   //  /System/Library/Frameworks/CoreServices.framework/Versions/A/CoreServices (compatibility version 1.0.0, current version 1228.0.0)
+  //  /System/Library/Frameworks/QuartzCore.framework/Versions/A/QuartzCore (compatibility version 1.2.0, current version 1.11.0, weak)
+  //  /System/Library/Frameworks/Metal.framework/Versions/A/Metal (compatibility version 1.0.0, current version 341.16.0, weak)
+  //  /System/Library/Frameworks/CoreGraphics.framework/Versions/A/CoreGraphics (compatibility version 64.0.0, current version 1774.0.4, weak)
+  //  /System/Library/Frameworks/MetalPerformanceShaders.framework/Versions/A/MetalPerformanceShaders (compatibility version 1.0.0, current version 127.0.19, weak)
   // 	/usr/lib/libiconv.2.dylib (compatibility version 7.0.0, current version 7.0.0)
   // 	/usr/lib/libSystem.B.dylib (compatibility version 1.0.0, current version 1319.0.0)
-  const EXPECTED: [&str; 5] = [
+  // 	/usr/lib/libobjc.A.dylib (compatibility version 1.0.0, current version 228.0.0)
-    "/System/Library/Frameworks/CoreFoundation.framework/Versions/A/CoreFoundation",
+
-    "/System/Library/Frameworks/CoreServices.framework/Versions/A/CoreServices",
+  // path and whether its weak or not
-    "/usr/lib/libiconv.2.dylib",
+  const EXPECTED: [(&str, bool); 9] = [
-    "/usr/lib/libSystem.B.dylib",
+    ("/System/Library/Frameworks/CoreFoundation.framework/Versions/A/CoreFoundation", false),
-    "/usr/lib/libobjc.A.dylib",
+    ("/System/Library/Frameworks/CoreServices.framework/Versions/A/CoreServices", false),
+    ("/System/Library/Frameworks/QuartzCore.framework/Versions/A/QuartzCore", true),
+    ("/System/Library/Frameworks/Metal.framework/Versions/A/Metal", true),
+    ("/System/Library/Frameworks/CoreGraphics.framework/Versions/A/CoreGraphics", true),
+    ("/System/Library/Frameworks/MetalPerformanceShaders.framework/Versions/A/MetalPerformanceShaders", true),
+    ("/usr/lib/libiconv.2.dylib", false),
+    ("/usr/lib/libSystem.B.dylib", false),
+    ("/usr/lib/libobjc.A.dylib", false),
   ];
 
   let otool = std::process::Command::new("otool")
@@ -64,9 +75,9 @@ fn macos_shared_libraries() {
   let output = std::str::from_utf8(&otool.stdout).unwrap();
   // Ensure that the output contains only the expected shared libraries.
   for line in output.lines().skip(1) {
-    let path = line.split_whitespace().next().unwrap();
+    let (path, attributes) = line.trim().split_once(' ').unwrap();
     assert!(
-      EXPECTED.contains(&path),
+      EXPECTED.contains(&(path, attributes.ends_with("weak)"))),
       "Unexpected shared library: {}",
       path
     );

cli/tests/testdata/run/unstable_webgpu.disabled.out

@@ -0,0 +1,2 @@
+main undefined
+worker undefined

cli/tests/testdata/run/unstable_webgpu.enabled.out

@@ -0,0 +1,2 @@
+main [class GPU]
+worker [class GPU]

cli/tests/testdata/run/unstable_webgpu.js

@@ -0,0 +1,10 @@
+const scope = import.meta.url.slice(-7) === "#worker" ? "worker" : "main";
+
+console.log(scope, globalThis.GPU);
+
+if (scope === "worker") {
+  postMessage("done");
+} else {
+  const worker = new Worker(`${import.meta.url}#worker`, { type: "module" });
+  worker.onmessage = () => Deno.exit(0);
+}

cli/tests/testdata/webgpu/computepass_shader.wgsl

@@ -0,0 +1,38 @@
+@group(0)
+@binding(0)
+var<storage, read_write> v_indices: array<u32>; // this is used as both input and output for convenience
+
+// The Collatz Conjecture states that for any integer n:
+// If n is even, n = n/2
+// If n is odd, n = 3n+1
+// And repeat this process for each new n, you will always eventually reach 1.
+// Though the conjecture has not been proven, no counterexample has ever been found.
+// This function returns how many times this recurrence needs to be applied to reach 1.
+fn collatz_iterations(n_base: u32) -> u32{
+    var n: u32 = n_base;
+    var i: u32 = 0u;
+    loop {
+        if (n <= 1u) {
+            break;
+        }
+        if (n % 2u == 0u) {
+            n = n / 2u;
+        }
+        else {
+            // Overflow? (i.e. 3*n + 1 > 0xffffffffu?)
+            if (n >= 1431655765u) {   // 0x55555555u
+                return 4294967295u;   // 0xffffffffu
+            }
+
+            n = 3u * n + 1u;
+        }
+        i = i + 1u;
+    }
+    return i;
+}
+
+@compute
+@workgroup_size(1)
+fn main(@builtin(global_invocation_id) global_id: vec3<u32>) {
+    v_indices[global_id.x] = collatz_iterations(v_indices[global_id.x]);
+}

cli/tests/testdata/webgpu/hellotriangle_shader.wgsl

@@ -0,0 +1,11 @@
+@vertex
+fn vs_main(@builtin(vertex_index) in_vertex_index: u32) -> @builtin(position) vec4<f32> {
+    let x = f32(i32(in_vertex_index) - 1);
+    let y = f32(i32(in_vertex_index & 1u) * 2 - 1);
+    return vec4<f32>(x, y, 0.0, 1.0);
+}
+
+@fragment
+fn fs_main() -> @location(0) vec4<f32> {
+    return vec4<f32>(1.0, 0.0, 0.0, 1.0);
+}

cli/tests/unit/webgpu_test.ts

@@ -0,0 +1,242 @@
+// Copyright 2018-2023 the Deno authors. All rights reserved. MIT license.
+
+import { assert, assertEquals } from "./test_util.ts";
+
+let isCI: boolean;
+try {
+  isCI = (Deno.env.get("CI")?.length ?? 0) > 0;
+} catch {
+  isCI = true;
+}
+
+// Skip these tests on linux CI, because the vulkan emulator is not good enough
+// yet, and skip on macOS CI because these do not have virtual GPUs.
+const isLinuxOrMacCI =
+  (Deno.build.os === "linux" || Deno.build.os === "darwin") && isCI;
+// Skip these tests in WSL because it doesn't have good GPU support.
+const isWsl = await checkIsWsl();
+
+Deno.test({
+  permissions: { read: true, env: true },
+  ignore: isWsl || isLinuxOrMacCI,
+}, async function webgpuComputePass() {
+  const adapter = await navigator.gpu.requestAdapter();
+  assert(adapter);
+
+  const numbers = [1, 4, 3, 295];
+
+  const device = await adapter.requestDevice();
+  assert(device);
+
+  const shaderCode = await Deno.readTextFile(
+    "cli/tests/testdata/webgpu/computepass_shader.wgsl",
+  );
+
+  const shaderModule = device.createShaderModule({
+    code: shaderCode,
+  });
+
+  const size = new Uint32Array(numbers).byteLength;
+
+  const stagingBuffer = device.createBuffer({
+    size: size,
+    usage: GPUBufferUsage.MAP_READ | GPUBufferUsage.COPY_DST,
+  });
+
+  const storageBuffer = device.createBuffer({
+    label: "Storage Buffer",
+    size: size,
+    usage: GPUBufferUsage.STORAGE | GPUBufferUsage.COPY_DST |
+      GPUBufferUsage.COPY_SRC,
+    mappedAtCreation: true,
+  });
+
+  const buf = new Uint32Array(storageBuffer.getMappedRange());
+
+  buf.set(numbers);
+
+  storageBuffer.unmap();
+
+  const computePipeline = device.createComputePipeline({
+    layout: "auto",
+    compute: {
+      module: shaderModule,
+      entryPoint: "main",
+    },
+  });
+  const bindGroupLayout = computePipeline.getBindGroupLayout(0);
+
+  const bindGroup = device.createBindGroup({
+    layout: bindGroupLayout,
+    entries: [
+      {
+        binding: 0,
+        resource: {
+          buffer: storageBuffer,
+        },
+      },
+    ],
+  });
+
+  const encoder = device.createCommandEncoder();
+
+  const computePass = encoder.beginComputePass();
+  computePass.setPipeline(computePipeline);
+  computePass.setBindGroup(0, bindGroup);
+  computePass.insertDebugMarker("compute collatz iterations");
+  computePass.dispatchWorkgroups(numbers.length);
+  computePass.end();
+
+  encoder.copyBufferToBuffer(storageBuffer, 0, stagingBuffer, 0, size);
+
+  device.queue.submit([encoder.finish()]);
+
+  await stagingBuffer.mapAsync(1);
+
+  const data = stagingBuffer.getMappedRange();
+
+  assertEquals(new Uint32Array(data), new Uint32Array([0, 2, 7, 55]));
+
+  stagingBuffer.unmap();
+
+  device.destroy();
+
+  // TODO(lucacasonato): webgpu spec should add a explicit destroy method for
+  // adapters.
+  const resources = Object.keys(Deno.resources());
+  Deno.close(Number(resources[resources.length - 1]));
+});
+
+Deno.test({
+  permissions: { read: true, env: true },
+  ignore: isWsl || isLinuxOrMacCI,
+}, async function webgpuHelloTriangle() {
+  const adapter = await navigator.gpu.requestAdapter();
+  assert(adapter);
+
+  const device = await adapter.requestDevice();
+  assert(device);
+
+  const shaderCode = await Deno.readTextFile(
+    "cli/tests/testdata/webgpu/hellotriangle_shader.wgsl",
+  );
+
+  const shaderModule = device.createShaderModule({
+    code: shaderCode,
+  });
+
+  const pipelineLayout = device.createPipelineLayout({
+    bindGroupLayouts: [],
+  });
+
+  const renderPipeline = device.createRenderPipeline({
+    layout: pipelineLayout,
+    vertex: {
+      module: shaderModule,
+      entryPoint: "vs_main",
+    },
+    fragment: {
+      module: shaderModule,
+      entryPoint: "fs_main",
+      targets: [
+        {
+          format: "rgba8unorm-srgb",
+        },
+      ],
+    },
+  });
+
+  const dimensions = {
+    width: 200,
+    height: 200,
+  };
+  const unpaddedBytesPerRow = dimensions.width * 4;
+  const align = 256;
+  const paddedBytesPerRowPadding = (align - unpaddedBytesPerRow % align) %
+    align;
+  const paddedBytesPerRow = unpaddedBytesPerRow + paddedBytesPerRowPadding;
+
+  const outputBuffer = device.createBuffer({
+    label: "Capture",
+    size: paddedBytesPerRow * dimensions.height,
+    usage: GPUBufferUsage.MAP_READ | GPUBufferUsage.COPY_DST,
+  });
+  const texture = device.createTexture({
+    label: "Capture",
+    size: dimensions,
+    format: "rgba8unorm-srgb",
+    usage: GPUTextureUsage.RENDER_ATTACHMENT | GPUTextureUsage.COPY_SRC,
+  });
+
+  const encoder = device.createCommandEncoder();
+  const view = texture.createView();
+  const renderPass = encoder.beginRenderPass({
+    colorAttachments: [
+      {
+        view,
+        storeOp: "store",
+        loadOp: "clear",
+        clearValue: [0, 1, 0, 1],
+      },
+    ],
+  });
+  renderPass.setPipeline(renderPipeline);
+  renderPass.draw(3, 1);
+  renderPass.end();
+
+  encoder.copyTextureToBuffer(
+    {
+      texture,
+    },
+    {
+      buffer: outputBuffer,
+      bytesPerRow: paddedBytesPerRow,
+      rowsPerImage: 0,
+    },
+    dimensions,
+  );
+
+  const bundle = encoder.finish();
+  device.queue.submit([bundle]);
+
+  await outputBuffer.mapAsync(1);
+  const data = new Uint8Array(outputBuffer.getMappedRange());
+
+  assertEquals(
+    data,
+    await Deno.readFile("cli/tests/testdata/webgpu/hellotriangle.out"),
+  );
+
+  outputBuffer.unmap();
+
+  device.destroy();
+
+  // TODO(lucacasonato): webgpu spec should add a explicit destroy method for
+  // adapters.
+  const resources = Object.keys(Deno.resources());
+  Deno.close(Number(resources[resources.length - 1]));
+});
+
+Deno.test({
+  ignore: isWsl || isLinuxOrMacCI,
+}, async function webgpuAdapterHasFeatures() {
+  const adapter = await navigator.gpu.requestAdapter();
+  assert(adapter);
+  assert(adapter.features);
+  const resources = Object.keys(Deno.resources());
+  Deno.close(Number(resources[resources.length - 1]));
+});
+
+async function checkIsWsl() {
+  return Deno.build.os === "linux" && await hasMicrosoftProcVersion();
+
+  async function hasMicrosoftProcVersion() {
+    // https://github.com/microsoft/WSL/issues/423#issuecomment-221627364
+    try {
+      const procVersion = await Deno.readTextFile("/proc/version");
+      return /microsoft/i.test(procVersion);
+    } catch {
+      return false;
+    }
+  }
+}

cli/tsc/dts/lib.deno.window.d.ts

@@ -3,6 +3,7 @@
 /// <reference no-default-lib="true" />
 /// <reference lib="deno.ns" />
 /// <reference lib="deno.shared_globals" />
+/// <reference lib="deno.webgpu" />
 /// <reference lib="deno.webstorage" />
 /// <reference lib="esnext" />
 /// <reference lib="deno.cache" />
@@ -102,6 +103,7 @@ declare var caches: CacheStorage;
 
 /** @category Web APIs */
 declare interface Navigator {
+  readonly gpu: GPU;
   readonly hardwareConcurrency: number;
   readonly userAgent: string;
   readonly language: string;

cli/tsc/dts/lib.deno.worker.d.ts

@@ -62,6 +62,7 @@ declare var WorkerGlobalScope: {
 
 /** @category Web APIs */
 declare interface WorkerNavigator {
+  readonly gpu: GPU;
   readonly hardwareConcurrency: number;
   readonly userAgent: string;
   readonly language: string;

cli/tsc/mod.rs

@@ -92,6 +92,7 @@ pub fn get_types_declaration_file_text(unstable: bool) -> String {
     "deno.url",
     "deno.web",
     "deno.fetch",
+    "deno.webgpu",
     "deno.websocket",
     "deno.webstorage",
     "deno.crypto",

ext/webgpu/02_surface.js

@@ -0,0 +1,235 @@
+// Copyright 2018-2023 the Deno authors. All rights reserved. MIT license.
+
+// @ts-check
+/// <reference path="../../core/lib.deno_core.d.ts" />
+/// <reference path="../web/internal.d.ts" />
+/// <reference path="../web/lib.deno_web.d.ts" />
+/// <reference path="./lib.deno_webgpu.d.ts" />
+
+import { core, primordials } from "ext:core/mod.js";
+const ops = core.ops;
+import * as webidl from "ext:deno_webidl/00_webidl.js";
+import { createFilteredInspectProxy } from "ext:deno_console/01_console.js";
+const { Symbol, SymbolFor, ObjectPrototypeIsPrototypeOf } = primordials;
+import {
+  _device,
+  assertDevice,
+  createGPUTexture,
+  GPUTextureUsage,
+} from "ext:deno_webgpu/01_webgpu.js";
+
+const _surfaceRid = Symbol("[[surfaceRid]]");
+const _configuration = Symbol("[[configuration]]");
+const _canvas = Symbol("[[canvas]]");
+const _currentTexture = Symbol("[[currentTexture]]");
+class GPUCanvasContext {
+  /** @type {number} */
+  [_surfaceRid];
+  /** @type {InnerGPUDevice} */
+  [_device];
+  [_configuration];
+  [_canvas];
+  /** @type {GPUTexture | undefined} */
+  [_currentTexture];
+
+  get canvas() {
+    webidl.assertBranded(this, GPUCanvasContextPrototype);
+    return this[_canvas];
+  }
+
+  constructor() {
+    webidl.illegalConstructor();
+  }
+
+  configure(configuration) {
+    webidl.assertBranded(this, GPUCanvasContextPrototype);
+    const prefix = "Failed to execute 'configure' on 'GPUCanvasContext'";
+    webidl.requiredArguments(arguments.length, 1, { prefix });
+    configuration = webidl.converters.GPUCanvasConfiguration(configuration, {
+      prefix,
+      context: "Argument 1",
+    });
+
+    this[_device] = configuration.device[_device];
+    this[_configuration] = configuration;
+    const device = assertDevice(this, {
+      prefix,
+      context: "configuration.device",
+    });
+
+    const { err } = ops.op_webgpu_surface_configure({
+      surfaceRid: this[_surfaceRid],
+      deviceRid: device.rid,
+      format: configuration.format,
+      viewFormats: configuration.viewFormats,
+      usage: configuration.usage,
+      width: configuration.width,
+      height: configuration.height,
+      alphaMode: configuration.alphaMode,
+    });
+
+    device.pushError(err);
+  }
+
+  unconfigure() {
+    webidl.assertBranded(this, GPUCanvasContextPrototype);
+
+    this[_configuration] = null;
+    this[_device] = null;
+  }
+
+  getCurrentTexture() {
+    webidl.assertBranded(this, GPUCanvasContextPrototype);
+    const prefix =
+      "Failed to execute 'getCurrentTexture' on 'GPUCanvasContext'";
+
+    if (this[_configuration] === null) {
+      throw new DOMException("context is not configured.", "InvalidStateError");
+    }
+
+    const device = assertDevice(this, { prefix, context: "this" });
+
+    if (this[_currentTexture]) {
+      return this[_currentTexture];
+    }
+
+    const { rid } = ops.op_webgpu_surface_get_current_texture(
+      device.rid,
+      this[_surfaceRid],
+    );
+
+    const texture = createGPUTexture(
+      {
+        size: {
+          width: this[_configuration].width,
+          height: this[_configuration].height,
+          depthOrArrayLayers: 1,
+        },
+        mipLevelCount: 1,
+        sampleCount: 1,
+        dimension: "2d",
+        format: this[_configuration].format,
+        usage: this[_configuration].usage,
+      },
+      device,
+      rid,
+    );
+    device.trackResource(texture);
+    this[_currentTexture] = texture;
+    return texture;
+  }
+
+  // Extended from spec. Required to present the texture; browser don't need this.
+  present() {
+    webidl.assertBranded(this, GPUCanvasContextPrototype);
+    const prefix = "Failed to execute 'present' on 'GPUCanvasContext'";
+    const device = assertDevice(this[_currentTexture], {
+      prefix,
+      context: "this",
+    });
+    ops.op_webgpu_surface_present(device.rid, this[_surfaceRid]);
+    this[_currentTexture].destroy();
+    this[_currentTexture] = undefined;
+  }
+
+  [SymbolFor("Deno.privateCustomInspect")](inspect, inspectOptions) {
+    return inspect(
+      createFilteredInspectProxy({
+        object: this,
+        evaluate: ObjectPrototypeIsPrototypeOf(GPUCanvasContextPrototype, this),
+        keys: [
+          "canvas",
+        ],
+      }),
+      inspectOptions,
+    );
+  }
+}
+const GPUCanvasContextPrototype = GPUCanvasContext.prototype;
+
+function createCanvasContext(options) {
+  const canvasContext = webidl.createBranded(GPUCanvasContext);
+  canvasContext[_surfaceRid] = options.surfaceRid;
+  canvasContext[_canvas] = options.canvas;
+  return canvasContext;
+}
+
+// Converters
+
+// ENUM: GPUCanvasAlphaMode
+webidl.converters["GPUCanvasAlphaMode"] = webidl.createEnumConverter(
+  "GPUCanvasAlphaMode",
+  [
+    "opaque",
+    "premultiplied",
+  ],
+);
+
+// NON-SPEC: ENUM: GPUPresentMode
+webidl.converters["GPUPresentMode"] = webidl.createEnumConverter(
+  "GPUPresentMode",
+  [
+    "autoVsync",
+    "autoNoVsync",
+    "fifo",
+    "fifoRelaxed",
+    "immediate",
+    "mailbox",
+  ],
+);
+
+// DICT: GPUCanvasConfiguration
+const dictMembersGPUCanvasConfiguration = [
+  { key: "device", converter: webidl.converters.GPUDevice, required: true },
+  {
+    key: "format",
+    converter: webidl.converters.GPUTextureFormat,
+    required: true,
+  },
+  {
+    key: "usage",
+    converter: webidl.converters["GPUTextureUsageFlags"],
+    defaultValue: GPUTextureUsage.RENDER_ATTACHMENT,
+  },
+  {
+    key: "alphaMode",
+    converter: webidl.converters["GPUCanvasAlphaMode"],
+    defaultValue: "opaque",
+  },
+
+  // Extended from spec
+  {
+    key: "presentMode",
+    converter: webidl.converters["GPUPresentMode"],
+  },
+  {
+    key: "width",
+    converter: webidl.converters["long"],
+    required: true,
+  },
+  {
+    key: "height",
+    converter: webidl.converters["long"],
+    required: true,
+  },
+  {
+    key: "viewFormats",
+    converter: webidl.createSequenceConverter(
+      webidl.converters["GPUTextureFormat"],
+    ),
+    get defaultValue() {
+      return [];
+    },
+  },
+];
+webidl.converters["GPUCanvasConfiguration"] = webidl
+  .createDictionaryConverter(
+    "GPUCanvasConfiguration",
+    dictMembersGPUCanvasConfiguration,
+  );
+
+window.__bootstrap.webgpu = {
+  ...window.__bootstrap.webgpu,
+  GPUCanvasContext,
+  createCanvasContext,
+};

ext/webgpu/Cargo.toml

@@ -0,0 +1,49 @@
+# Copyright 2018-2023 the Deno authors. All rights reserved. MIT license.
+
+[package]
+name = "deno_webgpu"
+version = "0.94.0"
+authors = ["the Deno authors"]
+edition.workspace = true
+license = "MIT"
+readme = "README.md"
+repository = "https://github.com/gfx-rs/wgpu"
+description = "WebGPU implementation for Deno"
+
+[lib]
+path = "lib.rs"
+
+[features]
+surface = ["wgpu-core/raw-window-handle", "dep:raw-window-handle"]
+
+# We make all dependencies conditional on not being wasm,
+# so the whole workspace can built as wasm.
+[target.'cfg(not(target_arch = "wasm32"))'.dependencies]
+deno_core.workspace = true
+serde = { workspace = true, features = ["derive"] }
+tokio = { workspace = true, features = ["full"] }
+wgpu-types = { workspace = true, features = ["trace", "replay", "serde"] }
+raw-window-handle = { workspace = true, optional = true }
+
+[target.'cfg(not(target_arch = "wasm32"))'.dependencies.wgpu-core]
+workspace = true
+features = ["trace", "replay", "serde", "strict_asserts", "wgsl", "gles"]
+
+# We want the wgpu-core Metal backend on macOS and iOS.
+[target.'cfg(any(target_os = "macos", target_os = "ios"))'.dependencies.wgpu-core]
+workspace = true
+features = ["metal"]
+
+# We want the wgpu-core Direct3D backends on Windows.
+[target.'cfg(windows)'.dependencies.wgpu-core]
+workspace = true
+features = ["dx11", "dx12"]
+
+[target.'cfg(windows)'.dependencies.wgpu-hal]
+workspace = true
+features = ["windows_rs"]
+
+# We want the wgpu-core Vulkan backend on Unix (but not Emscripten) and Windows.
+[target.'cfg(any(windows, all(unix, not(target_os = "emscripten"))))'.dependencies.wgpu-core]
+workspace = true
+features = ["vulkan"]

ext/webgpu/LICENSE.md

@@ -0,0 +1,20 @@
+MIT License
+
+Copyright 2018-2023 the Deno authors
+
+Permission is hereby granted, free of charge, to any person obtaining a copy of
+this software and associated documentation files (the "Software"), to deal in
+the Software without restriction, including without limitation the rights to
+use, copy, modify, merge, publish, distribute, sublicense, and/or sell copies of
+the Software, and to permit persons to whom the Software is furnished to do so,
+subject to the following conditions:
+
+The above copyright notice and this permission notice shall be included in all
+copies or substantial portions of the Software.
+
+THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
+IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, FITNESS
+FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE AUTHORS OR
+COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER LIABILITY, WHETHER
+IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, OUT OF OR IN
+CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE SOFTWARE.

ext/webgpu/README.md

@@ -0,0 +1,35 @@
+# deno_webgpu
+
+This op crate implements the WebGPU API as defined in
+https://gpuweb.github.io/gpuweb/ in Deno. The implementation targets the spec
+draft as of October 4, 2023. The spec is still very much in flux. This op crate
+tries to stay up to date with the spec, but is constrained by the features
+implemented in our GPU backend library [wgpu](https://github.com/gfx-rs/wgpu).
+
+The spec is still very bare bones, and is still missing many details. As the
+spec becomes more concrete, we will implement to follow the spec more closely.
+
+In addition, setting the `DENO_WEBGPU_TRACE` environmental variable will output
+a
+[wgpu trace](https://github.com/gfx-rs/wgpu/wiki/Debugging-wgpu-Applications#tracing-infrastructure)
+to the specified directory.
+
+For testing this op crate will make use of the WebGPU conformance tests suite,
+running through our WPT runner. This will be used to validate implementation
+conformance.
+
+GitHub CI doesn't run with GPUs, so testing relies on software like DX WARP &
+Vulkan lavapipe. Currently only using DX WARP works, so tests are only run on
+Windows.
+
+## Links
+
+Specification: https://gpuweb.github.io/gpuweb/
+
+Design documents: https://github.com/gpuweb/gpuweb/tree/main/design
+
+Conformance tests suite: https://github.com/gpuweb/cts
+
+WebGPU examples for Deno: https://github.com/crowlKats/webgpu-examples
+
+wgpu-users matrix channel: https://matrix.to/#/#wgpu-users:matrix.org

ext/webgpu/binding.rs

@@ -0,0 +1,340 @@
+// Copyright 2018-2023 the Deno authors. All rights reserved. MIT license.
+
+use deno_core::error::AnyError;
+use deno_core::op2;
+use deno_core::OpState;
+use deno_core::Resource;
+use deno_core::ResourceId;
+use serde::Deserialize;
+use std::borrow::Cow;
+use std::rc::Rc;
+
+use super::error::WebGpuResult;
+
+pub(crate) struct WebGpuBindGroupLayout(
+  pub(crate) crate::Instance,
+  pub(crate) wgpu_core::id::BindGroupLayoutId,
+);
+impl Resource for WebGpuBindGroupLayout {
+  fn name(&self) -> Cow<str> {
+    "webGPUBindGroupLayout".into()
+  }
+
+  fn close(self: Rc<Self>) {
+    let instance = &self.0;
+    gfx_select!(self.1 => instance.bind_group_layout_drop(self.1));
+  }
+}
+
+pub(crate) struct WebGpuBindGroup(
+  pub(crate) crate::Instance,
+  pub(crate) wgpu_core::id::BindGroupId,
+);
+impl Resource for WebGpuBindGroup {
+  fn name(&self) -> Cow<str> {
+    "webGPUBindGroup".into()
+  }
+
+  fn close(self: Rc<Self>) {
+    let instance = &self.0;
+    gfx_select!(self.1 => instance.bind_group_drop(self.1));
+  }
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "camelCase")]
+struct GpuBufferBindingLayout {
+  r#type: GpuBufferBindingType,
+  has_dynamic_offset: bool,
+  min_binding_size: u64,
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "kebab-case")]
+enum GpuBufferBindingType {
+  Uniform,
+  Storage,
+  ReadOnlyStorage,
+}
+
+impl From<GpuBufferBindingType> for wgpu_types::BufferBindingType {
+  fn from(binding_type: GpuBufferBindingType) -> Self {
+    match binding_type {
+      GpuBufferBindingType::Uniform => wgpu_types::BufferBindingType::Uniform,
+      GpuBufferBindingType::Storage => {
+        wgpu_types::BufferBindingType::Storage { read_only: false }
+      }
+      GpuBufferBindingType::ReadOnlyStorage => {
+        wgpu_types::BufferBindingType::Storage { read_only: true }
+      }
+    }
+  }
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "camelCase")]
+struct GpuSamplerBindingLayout {
+  r#type: wgpu_types::SamplerBindingType,
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "camelCase")]
+struct GpuTextureBindingLayout {
+  sample_type: GpuTextureSampleType,
+  view_dimension: wgpu_types::TextureViewDimension,
+  multisampled: bool,
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "kebab-case")]
+enum GpuTextureSampleType {
+  Float,
+  UnfilterableFloat,
+  Depth,
+  Sint,
+  Uint,
+}
+
+impl From<GpuTextureSampleType> for wgpu_types::TextureSampleType {
+  fn from(sample_type: GpuTextureSampleType) -> Self {
+    match sample_type {
+      GpuTextureSampleType::Float => {
+        wgpu_types::TextureSampleType::Float { filterable: true }
+      }
+      GpuTextureSampleType::UnfilterableFloat => {
+        wgpu_types::TextureSampleType::Float { filterable: false }
+      }
+      GpuTextureSampleType::Depth => wgpu_types::TextureSampleType::Depth,
+      GpuTextureSampleType::Sint => wgpu_types::TextureSampleType::Sint,
+      GpuTextureSampleType::Uint => wgpu_types::TextureSampleType::Uint,
+    }
+  }
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "camelCase")]
+struct GpuStorageTextureBindingLayout {
+  access: GpuStorageTextureAccess,
+  format: wgpu_types::TextureFormat,
+  view_dimension: wgpu_types::TextureViewDimension,
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "kebab-case")]
+enum GpuStorageTextureAccess {
+  WriteOnly,
+}
+
+impl From<GpuStorageTextureAccess> for wgpu_types::StorageTextureAccess {
+  fn from(access: GpuStorageTextureAccess) -> Self {
+    match access {
+      GpuStorageTextureAccess::WriteOnly => {
+        wgpu_types::StorageTextureAccess::WriteOnly
+      }
+    }
+  }
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "camelCase")]
+pub struct GpuBindGroupLayoutEntry {
+  binding: u32,
+  visibility: u32,
+  #[serde(flatten)]
+  binding_type: GpuBindingType,
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "camelCase")]
+enum GpuBindingType {
+  Buffer(GpuBufferBindingLayout),
+  Sampler(GpuSamplerBindingLayout),
+  Texture(GpuTextureBindingLayout),
+  StorageTexture(GpuStorageTextureBindingLayout),
+}
+
+impl From<GpuBindingType> for wgpu_types::BindingType {
+  fn from(binding_type: GpuBindingType) -> wgpu_types::BindingType {
+    match binding_type {
+      GpuBindingType::Buffer(buffer) => wgpu_types::BindingType::Buffer {
+        ty: buffer.r#type.into(),
+        has_dynamic_offset: buffer.has_dynamic_offset,
+        min_binding_size: std::num::NonZeroU64::new(buffer.min_binding_size),
+      },
+      GpuBindingType::Sampler(sampler) => {
+        wgpu_types::BindingType::Sampler(sampler.r#type)
+      }
+      GpuBindingType::Texture(texture) => wgpu_types::BindingType::Texture {
+        sample_type: texture.sample_type.into(),
+        view_dimension: texture.view_dimension,
+        multisampled: texture.multisampled,
+      },
+      GpuBindingType::StorageTexture(storage_texture) => {
+        wgpu_types::BindingType::StorageTexture {
+          access: storage_texture.access.into(),
+          format: storage_texture.format,
+          view_dimension: storage_texture.view_dimension,
+        }
+      }
+    }
+  }
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_create_bind_group_layout(
+  state: &mut OpState,
+  #[smi] device_rid: ResourceId,
+  #[string] label: Cow<str>,
+  #[serde] entries: Vec<GpuBindGroupLayoutEntry>,
+) -> Result<WebGpuResult, AnyError> {
+  let instance = state.borrow::<super::Instance>();
+  let device_resource = state
+    .resource_table
+    .get::<super::WebGpuDevice>(device_rid)?;
+  let device = device_resource.1;
+
+  let entries = entries
+    .into_iter()
+    .map(|entry| {
+      wgpu_types::BindGroupLayoutEntry {
+        binding: entry.binding,
+        visibility: wgpu_types::ShaderStages::from_bits(entry.visibility)
+          .unwrap(),
+        ty: entry.binding_type.into(),
+        count: None, // native-only
+      }
+    })
+    .collect::<Vec<_>>();
+
+  let descriptor = wgpu_core::binding_model::BindGroupLayoutDescriptor {
+    label: Some(label),
+    entries: Cow::from(entries),
+  };
+
+  gfx_put!(device => instance.device_create_bind_group_layout(
+    device,
+    &descriptor,
+    ()
+  ) => state, WebGpuBindGroupLayout)
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_create_pipeline_layout(
+  state: &mut OpState,
+  #[smi] device_rid: ResourceId,
+  #[string] label: Cow<str>,
+  #[serde] bind_group_layouts: Vec<u32>,
+) -> Result<WebGpuResult, AnyError> {
+  let instance = state.borrow::<super::Instance>();
+  let device_resource = state
+    .resource_table
+    .get::<super::WebGpuDevice>(device_rid)?;
+  let device = device_resource.1;
+
+  let bind_group_layouts = bind_group_layouts
+    .into_iter()
+    .map(|rid| {
+      let bind_group_layout =
+        state.resource_table.get::<WebGpuBindGroupLayout>(rid)?;
+      Ok(bind_group_layout.1)
+    })
+    .collect::<Result<Vec<_>, AnyError>>()?;
+
+  let descriptor = wgpu_core::binding_model::PipelineLayoutDescriptor {
+    label: Some(label),
+    bind_group_layouts: Cow::from(bind_group_layouts),
+    push_constant_ranges: Default::default(),
+  };
+
+  gfx_put!(device => instance.device_create_pipeline_layout(
+    device,
+    &descriptor,
+    ()
+  ) => state, super::pipeline::WebGpuPipelineLayout)
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "camelCase")]
+pub struct GpuBindGroupEntry {
+  binding: u32,
+  kind: String,
+  resource: ResourceId,
+  offset: Option<u64>,
+  size: Option<u64>,
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_create_bind_group(
+  state: &mut OpState,
+  #[smi] device_rid: ResourceId,
+  #[string] label: Cow<str>,
+  #[smi] layout: ResourceId,
+  #[serde] entries: Vec<GpuBindGroupEntry>,
+) -> Result<WebGpuResult, AnyError> {
+  let instance = state.borrow::<super::Instance>();
+  let device_resource = state
+    .resource_table
+    .get::<super::WebGpuDevice>(device_rid)?;
+  let device = device_resource.1;
+
+  let entries = entries
+    .into_iter()
+    .map(|entry| {
+      Ok(wgpu_core::binding_model::BindGroupEntry {
+        binding: entry.binding,
+        resource: match entry.kind.as_str() {
+          "GPUSampler" => {
+            let sampler_resource =
+              state
+                .resource_table
+                .get::<super::sampler::WebGpuSampler>(entry.resource)?;
+            wgpu_core::binding_model::BindingResource::Sampler(
+              sampler_resource.1,
+            )
+          }
+          "GPUTextureView" => {
+            let texture_view_resource =
+              state
+                .resource_table
+                .get::<super::texture::WebGpuTextureView>(entry.resource)?;
+            wgpu_core::binding_model::BindingResource::TextureView(
+              texture_view_resource.1,
+            )
+          }
+          "GPUBufferBinding" => {
+            let buffer_resource =
+              state
+                .resource_table
+                .get::<super::buffer::WebGpuBuffer>(entry.resource)?;
+            wgpu_core::binding_model::BindingResource::Buffer(
+              wgpu_core::binding_model::BufferBinding {
+                buffer_id: buffer_resource.1,
+                offset: entry.offset.unwrap_or(0),
+                size: std::num::NonZeroU64::new(entry.size.unwrap_or(0)),
+              },
+            )
+          }
+          _ => unreachable!(),
+        },
+      })
+    })
+    .collect::<Result<Vec<_>, AnyError>>()?;
+
+  let bind_group_layout =
+    state.resource_table.get::<WebGpuBindGroupLayout>(layout)?;
+
+  let descriptor = wgpu_core::binding_model::BindGroupDescriptor {
+    label: Some(label),
+    layout: bind_group_layout.1,
+    entries: Cow::from(entries),
+  };
+
+  gfx_put!(device => instance.device_create_bind_group(
+    device,
+    &descriptor,
+    ()
+  ) => state, WebGpuBindGroup)
+}

ext/webgpu/buffer.rs

@@ -0,0 +1,205 @@
+// Copyright 2018-2023 the Deno authors. All rights reserved. MIT license.
+
+use deno_core::error::type_error;
+use deno_core::error::AnyError;
+use deno_core::futures::channel::oneshot;
+use deno_core::op2;
+use deno_core::OpState;
+use deno_core::Resource;
+use deno_core::ResourceId;
+use std::borrow::Cow;
+use std::cell::RefCell;
+use std::rc::Rc;
+use std::time::Duration;
+use wgpu_core::resource::BufferAccessResult;
+
+use super::error::DomExceptionOperationError;
+use super::error::WebGpuResult;
+
+pub(crate) struct WebGpuBuffer(
+  pub(crate) super::Instance,
+  pub(crate) wgpu_core::id::BufferId,
+);
+impl Resource for WebGpuBuffer {
+  fn name(&self) -> Cow<str> {
+    "webGPUBuffer".into()
+  }
+
+  fn close(self: Rc<Self>) {
+    let instance = &self.0;
+    gfx_select!(self.1 => instance.buffer_drop(self.1, true));
+  }
+}
+
+struct WebGpuBufferMapped(*mut u8, usize);
+impl Resource for WebGpuBufferMapped {
+  fn name(&self) -> Cow<str> {
+    "webGPUBufferMapped".into()
+  }
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_create_buffer(
+  state: &mut OpState,
+  #[smi] device_rid: ResourceId,
+  #[string] label: Cow<str>,
+  #[number] size: u64,
+  usage: u32,
+  mapped_at_creation: bool,
+) -> Result<WebGpuResult, AnyError> {
+  let instance = state.borrow::<super::Instance>();
+  let device_resource = state
+    .resource_table
+    .get::<super::WebGpuDevice>(device_rid)?;
+  let device = device_resource.1;
+
+  let descriptor = wgpu_core::resource::BufferDescriptor {
+    label: Some(label),
+    size,
+    usage: wgpu_types::BufferUsages::from_bits(usage)
+      .ok_or_else(|| type_error("usage is not valid"))?,
+    mapped_at_creation,
+  };
+
+  gfx_put!(device => instance.device_create_buffer(
+    device,
+    &descriptor,
+    ()
+  ) => state, WebGpuBuffer)
+}
+
+#[op2(async)]
+#[serde]
+pub async fn op_webgpu_buffer_get_map_async(
+  state: Rc<RefCell<OpState>>,
+  #[smi] buffer_rid: ResourceId,
+  #[smi] device_rid: ResourceId,
+  mode: u32,
+  #[number] offset: u64,
+  #[number] size: u64,
+) -> Result<WebGpuResult, AnyError> {
+  let (sender, receiver) = oneshot::channel::<BufferAccessResult>();
+
+  let device;
+  {
+    let state_ = state.borrow();
+    let instance = state_.borrow::<super::Instance>();
+    let buffer_resource =
+      state_.resource_table.get::<WebGpuBuffer>(buffer_rid)?;
+    let buffer = buffer_resource.1;
+    let device_resource = state_
+      .resource_table
+      .get::<super::WebGpuDevice>(device_rid)?;
+    device = device_resource.1;
+
+    let callback = Box::new(move |status| {
+      sender.send(status).unwrap();
+    });
+
+    // TODO(lucacasonato): error handling
+    let maybe_err = gfx_select!(buffer => instance.buffer_map_async(
+            buffer,
+            offset..(offset + size),
+            wgpu_core::resource::BufferMapOperation {
+                host: match mode {
+                    1 => wgpu_core::device::HostMap::Read,
+                    2 => wgpu_core::device::HostMap::Write,
+                    _ => unreachable!(),
+                },
+                callback: wgpu_core::resource::BufferMapCallback::from_rust(callback),
+            }
+        ))
+        .err();
+
+    if maybe_err.is_some() {
+      return Ok(WebGpuResult::maybe_err(maybe_err));
+    }
+  }
+
+  let done = Rc::new(RefCell::new(false));
+  let done_ = done.clone();
+  let device_poll_fut = async move {
+    while !*done.borrow() {
+      {
+        let state = state.borrow();
+        let instance = state.borrow::<super::Instance>();
+        gfx_select!(device => instance.device_poll(device, wgpu_types::Maintain::Wait))
+                    .unwrap();
+      }
+      tokio::time::sleep(Duration::from_millis(10)).await;
+    }
+    Ok::<(), AnyError>(())
+  };
+
+  let receiver_fut = async move {
+    receiver.await??;
+    let mut done = done_.borrow_mut();
+    *done = true;
+    Ok::<(), AnyError>(())
+  };
+
+  tokio::try_join!(device_poll_fut, receiver_fut)?;
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_buffer_get_mapped_range(
+  state: &mut OpState,
+  #[smi] buffer_rid: ResourceId,
+  #[number] offset: u64,
+  #[number] size: Option<u64>,
+  #[buffer] buf: &mut [u8],
+) -> Result<WebGpuResult, AnyError> {
+  let instance = state.borrow::<super::Instance>();
+  let buffer_resource = state.resource_table.get::<WebGpuBuffer>(buffer_rid)?;
+  let buffer = buffer_resource.1;
+
+  let (slice_pointer, range_size) =
+    gfx_select!(buffer => instance.buffer_get_mapped_range(
+      buffer,
+      offset,
+      size
+    ))
+    .map_err(|e| DomExceptionOperationError::new(&e.to_string()))?;
+
+  // SAFETY: guarantee to be safe from wgpu
+  let slice = unsafe {
+    std::slice::from_raw_parts_mut(slice_pointer, range_size as usize)
+  };
+  buf.copy_from_slice(slice);
+
+  let rid = state
+    .resource_table
+    .add(WebGpuBufferMapped(slice_pointer, range_size as usize));
+
+  Ok(WebGpuResult::rid(rid))
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_buffer_unmap(
+  state: &mut OpState,
+  #[smi] buffer_rid: ResourceId,
+  #[smi] mapped_rid: ResourceId,
+  #[buffer] buf: Option<&[u8]>,
+) -> Result<WebGpuResult, AnyError> {
+  let mapped_resource = state
+    .resource_table
+    .take::<WebGpuBufferMapped>(mapped_rid)?;
+  let instance = state.borrow::<super::Instance>();
+  let buffer_resource = state.resource_table.get::<WebGpuBuffer>(buffer_rid)?;
+  let buffer = buffer_resource.1;
+
+  if let Some(buf) = buf {
+    // SAFETY: guarantee to be safe from wgpu
+    let slice = unsafe {
+      std::slice::from_raw_parts_mut(mapped_resource.0, mapped_resource.1)
+    };
+    slice.copy_from_slice(buf);
+  }
+
+  gfx_ok!(buffer => instance.buffer_unmap(buffer))
+}

ext/webgpu/bundle.rs

@@ -0,0 +1,405 @@
+// Copyright 2018-2023 the Deno authors. All rights reserved. MIT license.
+
+use deno_core::error::type_error;
+use deno_core::error::AnyError;
+use deno_core::op2;
+use deno_core::OpState;
+use deno_core::Resource;
+use deno_core::ResourceId;
+use serde::Deserialize;
+use std::borrow::Cow;
+use std::cell::RefCell;
+use std::rc::Rc;
+
+use super::error::WebGpuResult;
+
+struct WebGpuRenderBundleEncoder(
+  RefCell<wgpu_core::command::RenderBundleEncoder>,
+);
+impl Resource for WebGpuRenderBundleEncoder {
+  fn name(&self) -> Cow<str> {
+    "webGPURenderBundleEncoder".into()
+  }
+}
+
+pub(crate) struct WebGpuRenderBundle(
+  pub(crate) super::Instance,
+  pub(crate) wgpu_core::id::RenderBundleId,
+);
+impl Resource for WebGpuRenderBundle {
+  fn name(&self) -> Cow<str> {
+    "webGPURenderBundle".into()
+  }
+
+  fn close(self: Rc<Self>) {
+    let instance = &self.0;
+    gfx_select!(self.1 => instance.render_bundle_drop(self.1));
+  }
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "camelCase")]
+pub struct CreateRenderBundleEncoderArgs {
+  device_rid: ResourceId,
+  label: String,
+  color_formats: Vec<Option<wgpu_types::TextureFormat>>,
+  depth_stencil_format: Option<wgpu_types::TextureFormat>,
+  sample_count: u32,
+  depth_read_only: bool,
+  stencil_read_only: bool,
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_create_render_bundle_encoder(
+  state: &mut OpState,
+  #[serde] args: CreateRenderBundleEncoderArgs,
+) -> Result<WebGpuResult, AnyError> {
+  let device_resource = state
+    .resource_table
+    .get::<super::WebGpuDevice>(args.device_rid)?;
+  let device = device_resource.1;
+
+  let depth_stencil = args.depth_stencil_format.map(|format| {
+    wgpu_types::RenderBundleDepthStencil {
+      format,
+      depth_read_only: args.depth_read_only,
+      stencil_read_only: args.stencil_read_only,
+    }
+  });
+
+  let descriptor = wgpu_core::command::RenderBundleEncoderDescriptor {
+    label: Some(Cow::Owned(args.label)),
+    color_formats: Cow::from(args.color_formats),
+    sample_count: args.sample_count,
+    depth_stencil,
+    multiview: None,
+  };
+
+  let res =
+    wgpu_core::command::RenderBundleEncoder::new(&descriptor, device, None);
+  let (render_bundle_encoder, maybe_err) = match res {
+    Ok(encoder) => (encoder, None),
+    Err(e) => (
+      wgpu_core::command::RenderBundleEncoder::dummy(device),
+      Some(e),
+    ),
+  };
+
+  let rid = state
+    .resource_table
+    .add(WebGpuRenderBundleEncoder(RefCell::new(
+      render_bundle_encoder,
+    )));
+
+  Ok(WebGpuResult::rid_err(rid, maybe_err))
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_render_bundle_encoder_finish(
+  state: &mut OpState,
+  #[smi] render_bundle_encoder_rid: ResourceId,
+  #[string] label: Cow<str>,
+) -> Result<WebGpuResult, AnyError> {
+  let render_bundle_encoder_resource =
+    state
+      .resource_table
+      .take::<WebGpuRenderBundleEncoder>(render_bundle_encoder_rid)?;
+  let render_bundle_encoder = Rc::try_unwrap(render_bundle_encoder_resource)
+    .ok()
+    .expect("unwrapping render_bundle_encoder_resource should succeed")
+    .0
+    .into_inner();
+  let instance = state.borrow::<super::Instance>();
+
+  gfx_put!(render_bundle_encoder.parent() => instance.render_bundle_encoder_finish(
+    render_bundle_encoder,
+    &wgpu_core::command::RenderBundleDescriptor {
+      label: Some(label),
+    },
+    ()
+  ) => state, WebGpuRenderBundle)
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_render_bundle_encoder_set_bind_group(
+  state: &mut OpState,
+  #[smi] render_bundle_encoder_rid: ResourceId,
+  index: u32,
+  #[smi] bind_group: ResourceId,
+  #[buffer] dynamic_offsets_data: &[u32],
+  #[number] dynamic_offsets_data_start: usize,
+  #[number] dynamic_offsets_data_length: usize,
+) -> Result<WebGpuResult, AnyError> {
+  let bind_group_resource =
+    state
+      .resource_table
+      .get::<super::binding::WebGpuBindGroup>(bind_group)?;
+  let render_bundle_encoder_resource =
+    state
+      .resource_table
+      .get::<WebGpuRenderBundleEncoder>(render_bundle_encoder_rid)?;
+
+  let start = dynamic_offsets_data_start;
+  let len = dynamic_offsets_data_length;
+
+  // Assert that length and start are both in bounds
+  assert!(start <= dynamic_offsets_data.len());
+  assert!(len <= dynamic_offsets_data.len() - start);
+
+  let dynamic_offsets_data = &dynamic_offsets_data[start..start + len];
+
+  // SAFETY: the raw pointer and length are of the same slice, and that slice
+  // lives longer than the below function invocation.
+  unsafe {
+    wgpu_core::command::bundle_ffi::wgpu_render_bundle_set_bind_group(
+      &mut render_bundle_encoder_resource.0.borrow_mut(),
+      index,
+      bind_group_resource.1,
+      dynamic_offsets_data.as_ptr(),
+      dynamic_offsets_data.len(),
+    );
+  }
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_render_bundle_encoder_push_debug_group(
+  state: &mut OpState,
+  #[smi] render_bundle_encoder_rid: ResourceId,
+  #[string] group_label: &str,
+) -> Result<WebGpuResult, AnyError> {
+  let render_bundle_encoder_resource =
+    state
+      .resource_table
+      .get::<WebGpuRenderBundleEncoder>(render_bundle_encoder_rid)?;
+
+  let label = std::ffi::CString::new(group_label).unwrap();
+  // SAFETY: the string the raw pointer points to lives longer than the below
+  // function invocation.
+  unsafe {
+    wgpu_core::command::bundle_ffi::wgpu_render_bundle_push_debug_group(
+      &mut render_bundle_encoder_resource.0.borrow_mut(),
+      label.as_ptr(),
+    );
+  }
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_render_bundle_encoder_pop_debug_group(
+  state: &mut OpState,
+  #[smi] render_bundle_encoder_rid: ResourceId,
+) -> Result<WebGpuResult, AnyError> {
+  let render_bundle_encoder_resource =
+    state
+      .resource_table
+      .get::<WebGpuRenderBundleEncoder>(render_bundle_encoder_rid)?;
+
+  wgpu_core::command::bundle_ffi::wgpu_render_bundle_pop_debug_group(
+    &mut render_bundle_encoder_resource.0.borrow_mut(),
+  );
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_render_bundle_encoder_insert_debug_marker(
+  state: &mut OpState,
+  #[smi] render_bundle_encoder_rid: ResourceId,
+  #[string] marker_label: &str,
+) -> Result<WebGpuResult, AnyError> {
+  let render_bundle_encoder_resource =
+    state
+      .resource_table
+      .get::<WebGpuRenderBundleEncoder>(render_bundle_encoder_rid)?;
+
+  let label = std::ffi::CString::new(marker_label).unwrap();
+  // SAFETY: the string the raw pointer points to lives longer than the below
+  // function invocation.
+  unsafe {
+    wgpu_core::command::bundle_ffi::wgpu_render_bundle_insert_debug_marker(
+      &mut render_bundle_encoder_resource.0.borrow_mut(),
+      label.as_ptr(),
+    );
+  }
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_render_bundle_encoder_set_pipeline(
+  state: &mut OpState,
+  #[smi] render_bundle_encoder_rid: ResourceId,
+  #[smi] pipeline: ResourceId,
+) -> Result<WebGpuResult, AnyError> {
+  let render_pipeline_resource =
+    state
+      .resource_table
+      .get::<super::pipeline::WebGpuRenderPipeline>(pipeline)?;
+  let render_bundle_encoder_resource =
+    state
+      .resource_table
+      .get::<WebGpuRenderBundleEncoder>(render_bundle_encoder_rid)?;
+
+  wgpu_core::command::bundle_ffi::wgpu_render_bundle_set_pipeline(
+    &mut render_bundle_encoder_resource.0.borrow_mut(),
+    render_pipeline_resource.1,
+  );
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_render_bundle_encoder_set_index_buffer(
+  state: &mut OpState,
+  #[smi] render_bundle_encoder_rid: ResourceId,
+  #[smi] buffer: ResourceId,
+  #[serde] index_format: wgpu_types::IndexFormat,
+  #[number] offset: u64,
+  #[number] size: u64,
+) -> Result<WebGpuResult, AnyError> {
+  let buffer_resource = state
+    .resource_table
+    .get::<super::buffer::WebGpuBuffer>(buffer)?;
+  let render_bundle_encoder_resource =
+    state
+      .resource_table
+      .get::<WebGpuRenderBundleEncoder>(render_bundle_encoder_rid)?;
+  let size = Some(
+    std::num::NonZeroU64::new(size)
+      .ok_or_else(|| type_error("size must be larger than 0"))?,
+  );
+
+  render_bundle_encoder_resource
+    .0
+    .borrow_mut()
+    .set_index_buffer(buffer_resource.1, index_format, offset, size);
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_render_bundle_encoder_set_vertex_buffer(
+  state: &mut OpState,
+  #[smi] render_bundle_encoder_rid: ResourceId,
+  slot: u32,
+  #[smi] buffer: ResourceId,
+  #[number] offset: u64,
+  #[number] size: Option<u64>,
+) -> Result<WebGpuResult, AnyError> {
+  let buffer_resource = state
+    .resource_table
+    .get::<super::buffer::WebGpuBuffer>(buffer)?;
+  let render_bundle_encoder_resource =
+    state
+      .resource_table
+      .get::<WebGpuRenderBundleEncoder>(render_bundle_encoder_rid)?;
+  let size = if let Some(size) = size {
+    Some(
+      std::num::NonZeroU64::new(size)
+        .ok_or_else(|| type_error("size must be larger than 0"))?,
+    )
+  } else {
+    None
+  };
+
+  wgpu_core::command::bundle_ffi::wgpu_render_bundle_set_vertex_buffer(
+    &mut render_bundle_encoder_resource.0.borrow_mut(),
+    slot,
+    buffer_resource.1,
+    offset,
+    size,
+  );
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_render_bundle_encoder_draw(
+  state: &mut OpState,
+  #[smi] render_bundle_encoder_rid: ResourceId,
+  vertex_count: u32,
+  instance_count: u32,
+  first_vertex: u32,
+  first_instance: u32,
+) -> Result<WebGpuResult, AnyError> {
+  let render_bundle_encoder_resource =
+    state
+      .resource_table
+      .get::<WebGpuRenderBundleEncoder>(render_bundle_encoder_rid)?;
+
+  wgpu_core::command::bundle_ffi::wgpu_render_bundle_draw(
+    &mut render_bundle_encoder_resource.0.borrow_mut(),
+    vertex_count,
+    instance_count,
+    first_vertex,
+    first_instance,
+  );
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_render_bundle_encoder_draw_indexed(
+  state: &mut OpState,
+  #[smi] render_bundle_encoder_rid: ResourceId,
+  index_count: u32,
+  instance_count: u32,
+  first_index: u32,
+  base_vertex: i32,
+  first_instance: u32,
+) -> Result<WebGpuResult, AnyError> {
+  let render_bundle_encoder_resource =
+    state
+      .resource_table
+      .get::<WebGpuRenderBundleEncoder>(render_bundle_encoder_rid)?;
+
+  wgpu_core::command::bundle_ffi::wgpu_render_bundle_draw_indexed(
+    &mut render_bundle_encoder_resource.0.borrow_mut(),
+    index_count,
+    instance_count,
+    first_index,
+    base_vertex,
+    first_instance,
+  );
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_render_bundle_encoder_draw_indirect(
+  state: &mut OpState,
+  #[smi] render_bundle_encoder_rid: ResourceId,
+  #[smi] indirect_buffer: ResourceId,
+  #[number] indirect_offset: u64,
+) -> Result<WebGpuResult, AnyError> {
+  let buffer_resource = state
+    .resource_table
+    .get::<super::buffer::WebGpuBuffer>(indirect_buffer)?;
+  let render_bundle_encoder_resource =
+    state
+      .resource_table
+      .get::<WebGpuRenderBundleEncoder>(render_bundle_encoder_rid)?;
+
+  wgpu_core::command::bundle_ffi::wgpu_render_bundle_draw_indirect(
+    &mut render_bundle_encoder_resource.0.borrow_mut(),
+    buffer_resource.1,
+    indirect_offset,
+  );
+
+  Ok(WebGpuResult::empty())
+}

ext/webgpu/command_encoder.rs

@@ -0,0 +1,633 @@
+// Copyright 2018-2023 the Deno authors. All rights reserved. MIT license.
+
+use crate::WebGpuQuerySet;
+use deno_core::error::AnyError;
+use deno_core::op2;
+use deno_core::OpState;
+use deno_core::Resource;
+use deno_core::ResourceId;
+use serde::Deserialize;
+use std::borrow::Cow;
+use std::cell::RefCell;
+use std::rc::Rc;
+
+use super::error::WebGpuResult;
+
+pub(crate) struct WebGpuCommandEncoder(
+  pub(crate) super::Instance,
+  pub(crate) wgpu_core::id::CommandEncoderId, // TODO: should maybe be option?
+);
+impl Resource for WebGpuCommandEncoder {
+  fn name(&self) -> Cow<str> {
+    "webGPUCommandEncoder".into()
+  }
+
+  fn close(self: Rc<Self>) {
+    let instance = &self.0;
+    gfx_select!(self.1 => instance.command_encoder_drop(self.1));
+  }
+}
+
+pub(crate) struct WebGpuCommandBuffer(
+  pub(crate) super::Instance,
+  pub(crate) RefCell<Option<wgpu_core::id::CommandBufferId>>,
+);
+impl Resource for WebGpuCommandBuffer {
+  fn name(&self) -> Cow<str> {
+    "webGPUCommandBuffer".into()
+  }
+
+  fn close(self: Rc<Self>) {
+    if let Some(id) = *self.1.borrow() {
+      let instance = &self.0;
+      gfx_select!(id => instance.command_buffer_drop(id));
+    }
+  }
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_create_command_encoder(
+  state: &mut OpState,
+  #[smi] device_rid: ResourceId,
+  #[string] label: Cow<str>,
+) -> Result<WebGpuResult, AnyError> {
+  let instance = state.borrow::<super::Instance>();
+  let device_resource = state
+    .resource_table
+    .get::<super::WebGpuDevice>(device_rid)?;
+  let device = device_resource.1;
+
+  let descriptor = wgpu_types::CommandEncoderDescriptor { label: Some(label) };
+
+  gfx_put!(device => instance.device_create_command_encoder(
+    device,
+    &descriptor,
+    ()
+  ) => state, WebGpuCommandEncoder)
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "camelCase")]
+pub struct GpuRenderPassColorAttachment {
+  view: ResourceId,
+  resolve_target: Option<ResourceId>,
+  clear_value: Option<wgpu_types::Color>,
+  load_op: wgpu_core::command::LoadOp,
+  store_op: wgpu_core::command::StoreOp,
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "camelCase")]
+pub struct GpuRenderPassDepthStencilAttachment {
+  view: ResourceId,
+  depth_clear_value: f32,
+  depth_load_op: Option<wgpu_core::command::LoadOp>,
+  depth_store_op: Option<wgpu_core::command::StoreOp>,
+  depth_read_only: bool,
+  stencil_clear_value: u32,
+  stencil_load_op: Option<wgpu_core::command::LoadOp>,
+  stencil_store_op: Option<wgpu_core::command::StoreOp>,
+  stencil_read_only: bool,
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "camelCase")]
+pub struct GPURenderPassTimestampWrites {
+  query_set: ResourceId,
+  beginning_of_pass_write_index: Option<u32>,
+  end_of_pass_write_index: Option<u32>,
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_command_encoder_begin_render_pass(
+  state: &mut OpState,
+  #[smi] command_encoder_rid: ResourceId,
+  #[string] label: Cow<str>,
+  #[serde] color_attachments: Vec<Option<GpuRenderPassColorAttachment>>,
+  #[serde] depth_stencil_attachment: Option<
+    GpuRenderPassDepthStencilAttachment,
+  >,
+  #[smi] occlusion_query_set: Option<ResourceId>,
+  #[serde] timestamp_writes: Option<GPURenderPassTimestampWrites>,
+) -> Result<WebGpuResult, AnyError> {
+  let command_encoder_resource = state
+    .resource_table
+    .get::<WebGpuCommandEncoder>(command_encoder_rid)?;
+
+  let color_attachments = color_attachments
+    .into_iter()
+    .map(|color_attachment| {
+      let rp_at = if let Some(at) = color_attachment.as_ref() {
+        let texture_view_resource =
+          state
+            .resource_table
+            .get::<super::texture::WebGpuTextureView>(at.view)?;
+
+        let resolve_target = at
+          .resolve_target
+          .map(|rid| {
+            state
+              .resource_table
+              .get::<super::texture::WebGpuTextureView>(rid)
+          })
+          .transpose()?
+          .map(|texture| texture.1);
+
+        Some(wgpu_core::command::RenderPassColorAttachment {
+          view: texture_view_resource.1,
+          resolve_target,
+          channel: wgpu_core::command::PassChannel {
+            load_op: at.load_op,
+            store_op: at.store_op,
+            clear_value: at.clear_value.unwrap_or_default(),
+            read_only: false,
+          },
+        })
+      } else {
+        None
+      };
+      Ok(rp_at)
+    })
+    .collect::<Result<Vec<_>, AnyError>>()?;
+
+  let mut processed_depth_stencil_attachment = None;
+
+  if let Some(attachment) = depth_stencil_attachment {
+    let texture_view_resource =
+      state
+        .resource_table
+        .get::<super::texture::WebGpuTextureView>(attachment.view)?;
+
+    processed_depth_stencil_attachment =
+      Some(wgpu_core::command::RenderPassDepthStencilAttachment {
+        view: texture_view_resource.1,
+        depth: wgpu_core::command::PassChannel {
+          load_op: attachment
+            .depth_load_op
+            .unwrap_or(wgpu_core::command::LoadOp::Load),
+          store_op: attachment
+            .depth_store_op
+            .unwrap_or(wgpu_core::command::StoreOp::Store),
+          clear_value: attachment.depth_clear_value,
+          read_only: attachment.depth_read_only,
+        },
+        stencil: wgpu_core::command::PassChannel {
+          load_op: attachment
+            .stencil_load_op
+            .unwrap_or(wgpu_core::command::LoadOp::Load),
+          store_op: attachment
+            .stencil_store_op
+            .unwrap_or(wgpu_core::command::StoreOp::Store),
+          clear_value: attachment.stencil_clear_value,
+          read_only: attachment.stencil_read_only,
+        },
+      });
+  }
+
+  let timestamp_writes = if let Some(timestamp_writes) = timestamp_writes {
+    let query_set_resource = state
+      .resource_table
+      .get::<WebGpuQuerySet>(timestamp_writes.query_set)?;
+    let query_set = query_set_resource.1;
+
+    Some(wgpu_core::command::RenderPassTimestampWrites {
+      query_set,
+      beginning_of_pass_write_index: timestamp_writes
+        .beginning_of_pass_write_index,
+      end_of_pass_write_index: timestamp_writes.end_of_pass_write_index,
+    })
+  } else {
+    None
+  };
+
+  let occlusion_query_set_resource = occlusion_query_set
+    .map(|rid| state.resource_table.get::<WebGpuQuerySet>(rid))
+    .transpose()?
+    .map(|query_set| query_set.1);
+
+  let descriptor = wgpu_core::command::RenderPassDescriptor {
+    label: Some(label),
+    color_attachments: Cow::from(color_attachments),
+    depth_stencil_attachment: processed_depth_stencil_attachment.as_ref(),
+    timestamp_writes: timestamp_writes.as_ref(),
+    occlusion_query_set: occlusion_query_set_resource,
+  };
+
+  let render_pass = wgpu_core::command::RenderPass::new(
+    command_encoder_resource.1,
+    &descriptor,
+  );
+
+  let rid = state
+    .resource_table
+    .add(super::render_pass::WebGpuRenderPass(RefCell::new(
+      render_pass,
+    )));
+
+  Ok(WebGpuResult::rid(rid))
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "camelCase")]
+pub struct GPUComputePassTimestampWrites {
+  query_set: ResourceId,
+  beginning_of_pass_write_index: Option<u32>,
+  end_of_pass_write_index: Option<u32>,
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_command_encoder_begin_compute_pass(
+  state: &mut OpState,
+  #[smi] command_encoder_rid: ResourceId,
+  #[string] label: Cow<str>,
+  #[serde] timestamp_writes: Option<GPUComputePassTimestampWrites>,
+) -> Result<WebGpuResult, AnyError> {
+  let command_encoder_resource = state
+    .resource_table
+    .get::<WebGpuCommandEncoder>(command_encoder_rid)?;
+
+  let timestamp_writes = if let Some(timestamp_writes) = timestamp_writes {
+    let query_set_resource = state
+      .resource_table
+      .get::<WebGpuQuerySet>(timestamp_writes.query_set)?;
+    let query_set = query_set_resource.1;
+
+    Some(wgpu_core::command::ComputePassTimestampWrites {
+      query_set,
+      beginning_of_pass_write_index: timestamp_writes
+        .beginning_of_pass_write_index,
+      end_of_pass_write_index: timestamp_writes.end_of_pass_write_index,
+    })
+  } else {
+    None
+  };
+
+  let descriptor = wgpu_core::command::ComputePassDescriptor {
+    label: Some(label),
+    timestamp_writes: timestamp_writes.as_ref(),
+  };
+
+  let compute_pass = wgpu_core::command::ComputePass::new(
+    command_encoder_resource.1,
+    &descriptor,
+  );
+
+  let rid = state
+    .resource_table
+    .add(super::compute_pass::WebGpuComputePass(RefCell::new(
+      compute_pass,
+    )));
+
+  Ok(WebGpuResult::rid(rid))
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_command_encoder_copy_buffer_to_buffer(
+  state: &mut OpState,
+  #[smi] command_encoder_rid: ResourceId,
+  #[smi] source: ResourceId,
+  #[number] source_offset: u64,
+  #[smi] destination: ResourceId,
+  #[number] destination_offset: u64,
+  #[number] size: u64,
+) -> Result<WebGpuResult, AnyError> {
+  let instance = state.borrow::<super::Instance>();
+  let command_encoder_resource = state
+    .resource_table
+    .get::<WebGpuCommandEncoder>(command_encoder_rid)?;
+  let command_encoder = command_encoder_resource.1;
+  let source_buffer_resource = state
+    .resource_table
+    .get::<super::buffer::WebGpuBuffer>(source)?;
+  let source_buffer = source_buffer_resource.1;
+  let destination_buffer_resource =
+    state
+      .resource_table
+      .get::<super::buffer::WebGpuBuffer>(destination)?;
+  let destination_buffer = destination_buffer_resource.1;
+
+  gfx_ok!(command_encoder => instance.command_encoder_copy_buffer_to_buffer(
+    command_encoder,
+    source_buffer,
+    source_offset,
+    destination_buffer,
+    destination_offset,
+    size
+  ))
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "camelCase")]
+pub struct GpuImageCopyBuffer {
+  buffer: ResourceId,
+  offset: u64,
+  bytes_per_row: Option<u32>,
+  rows_per_image: Option<u32>,
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "camelCase")]
+pub struct GpuImageCopyTexture {
+  pub texture: ResourceId,
+  pub mip_level: u32,
+  pub origin: wgpu_types::Origin3d,
+  pub aspect: wgpu_types::TextureAspect,
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_command_encoder_copy_buffer_to_texture(
+  state: &mut OpState,
+  #[smi] command_encoder_rid: ResourceId,
+  #[serde] source: GpuImageCopyBuffer,
+  #[serde] destination: GpuImageCopyTexture,
+  #[serde] copy_size: wgpu_types::Extent3d,
+) -> Result<WebGpuResult, AnyError> {
+  let instance = state.borrow::<super::Instance>();
+  let command_encoder_resource = state
+    .resource_table
+    .get::<WebGpuCommandEncoder>(command_encoder_rid)?;
+  let command_encoder = command_encoder_resource.1;
+  let source_buffer_resource =
+    state
+      .resource_table
+      .get::<super::buffer::WebGpuBuffer>(source.buffer)?;
+  let destination_texture_resource =
+    state
+      .resource_table
+      .get::<super::texture::WebGpuTexture>(destination.texture)?;
+
+  let source = wgpu_core::command::ImageCopyBuffer {
+    buffer: source_buffer_resource.1,
+    layout: wgpu_types::ImageDataLayout {
+      offset: source.offset,
+      bytes_per_row: source.bytes_per_row,
+      rows_per_image: source.rows_per_image,
+    },
+  };
+  let destination = wgpu_core::command::ImageCopyTexture {
+    texture: destination_texture_resource.id,
+    mip_level: destination.mip_level,
+    origin: destination.origin,
+    aspect: destination.aspect,
+  };
+  gfx_ok!(command_encoder => instance.command_encoder_copy_buffer_to_texture(
+    command_encoder,
+    &source,
+    &destination,
+    &copy_size
+  ))
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_command_encoder_copy_texture_to_buffer(
+  state: &mut OpState,
+  #[smi] command_encoder_rid: ResourceId,
+  #[serde] source: GpuImageCopyTexture,
+  #[serde] destination: GpuImageCopyBuffer,
+  #[serde] copy_size: wgpu_types::Extent3d,
+) -> Result<WebGpuResult, AnyError> {
+  let instance = state.borrow::<super::Instance>();
+  let command_encoder_resource = state
+    .resource_table
+    .get::<WebGpuCommandEncoder>(command_encoder_rid)?;
+  let command_encoder = command_encoder_resource.1;
+  let source_texture_resource =
+    state
+      .resource_table
+      .get::<super::texture::WebGpuTexture>(source.texture)?;
+  let destination_buffer_resource =
+    state
+      .resource_table
+      .get::<super::buffer::WebGpuBuffer>(destination.buffer)?;
+
+  let source = wgpu_core::command::ImageCopyTexture {
+    texture: source_texture_resource.id,
+    mip_level: source.mip_level,
+    origin: source.origin,
+    aspect: source.aspect,
+  };
+  let destination = wgpu_core::command::ImageCopyBuffer {
+    buffer: destination_buffer_resource.1,
+    layout: wgpu_types::ImageDataLayout {
+      offset: destination.offset,
+      bytes_per_row: destination.bytes_per_row,
+      rows_per_image: destination.rows_per_image,
+    },
+  };
+  gfx_ok!(command_encoder => instance.command_encoder_copy_texture_to_buffer(
+    command_encoder,
+    &source,
+    &destination,
+    &copy_size
+  ))
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_command_encoder_copy_texture_to_texture(
+  state: &mut OpState,
+  #[smi] command_encoder_rid: ResourceId,
+  #[serde] source: GpuImageCopyTexture,
+  #[serde] destination: GpuImageCopyTexture,
+  #[serde] copy_size: wgpu_types::Extent3d,
+) -> Result<WebGpuResult, AnyError> {
+  let instance = state.borrow::<super::Instance>();
+  let command_encoder_resource = state
+    .resource_table
+    .get::<WebGpuCommandEncoder>(command_encoder_rid)?;
+  let command_encoder = command_encoder_resource.1;
+  let source_texture_resource =
+    state
+      .resource_table
+      .get::<super::texture::WebGpuTexture>(source.texture)?;
+  let destination_texture_resource =
+    state
+      .resource_table
+      .get::<super::texture::WebGpuTexture>(destination.texture)?;
+
+  let source = wgpu_core::command::ImageCopyTexture {
+    texture: source_texture_resource.id,
+    mip_level: source.mip_level,
+    origin: source.origin,
+    aspect: source.aspect,
+  };
+  let destination = wgpu_core::command::ImageCopyTexture {
+    texture: destination_texture_resource.id,
+    mip_level: destination.mip_level,
+    origin: destination.origin,
+    aspect: destination.aspect,
+  };
+  gfx_ok!(command_encoder => instance.command_encoder_copy_texture_to_texture(
+    command_encoder,
+    &source,
+    &destination,
+    &copy_size
+  ))
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_command_encoder_clear_buffer(
+  state: &mut OpState,
+  #[smi] command_encoder_rid: ResourceId,
+  #[smi] buffer_rid: ResourceId,
+  #[number] offset: u64,
+  #[number] size: u64,
+) -> Result<WebGpuResult, AnyError> {
+  let instance = state.borrow::<super::Instance>();
+  let command_encoder_resource = state
+    .resource_table
+    .get::<WebGpuCommandEncoder>(command_encoder_rid)?;
+  let command_encoder = command_encoder_resource.1;
+  let destination_resource = state
+    .resource_table
+    .get::<super::buffer::WebGpuBuffer>(buffer_rid)?;
+
+  gfx_ok!(command_encoder => instance.command_encoder_clear_buffer(
+    command_encoder,
+    destination_resource.1,
+    offset,
+    std::num::NonZeroU64::new(size)
+  ))
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_command_encoder_push_debug_group(
+  state: &mut OpState,
+  #[smi] command_encoder_rid: ResourceId,
+  #[string] group_label: &str,
+) -> Result<WebGpuResult, AnyError> {
+  let instance = state.borrow::<super::Instance>();
+  let command_encoder_resource = state
+    .resource_table
+    .get::<WebGpuCommandEncoder>(command_encoder_rid)?;
+  let command_encoder = command_encoder_resource.1;
+
+  gfx_ok!(command_encoder => instance.command_encoder_push_debug_group(command_encoder, group_label))
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_command_encoder_pop_debug_group(
+  state: &mut OpState,
+  #[smi] command_encoder_rid: ResourceId,
+) -> Result<WebGpuResult, AnyError> {
+  let instance = state.borrow::<super::Instance>();
+  let command_encoder_resource = state
+    .resource_table
+    .get::<WebGpuCommandEncoder>(command_encoder_rid)?;
+  let command_encoder = command_encoder_resource.1;
+
+  gfx_ok!(command_encoder => instance.command_encoder_pop_debug_group(command_encoder))
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_command_encoder_insert_debug_marker(
+  state: &mut OpState,
+  #[smi] command_encoder_rid: ResourceId,
+  #[string] marker_label: &str,
+) -> Result<WebGpuResult, AnyError> {
+  let instance = state.borrow::<super::Instance>();
+  let command_encoder_resource = state
+    .resource_table
+    .get::<WebGpuCommandEncoder>(command_encoder_rid)?;
+  let command_encoder = command_encoder_resource.1;
+
+  gfx_ok!(command_encoder => instance.command_encoder_insert_debug_marker(
+    command_encoder,
+    marker_label
+  ))
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_command_encoder_write_timestamp(
+  state: &mut OpState,
+  #[smi] command_encoder_rid: ResourceId,
+  #[smi] query_set: ResourceId,
+  query_index: u32,
+) -> Result<WebGpuResult, AnyError> {
+  let instance = state.borrow::<super::Instance>();
+  let command_encoder_resource = state
+    .resource_table
+    .get::<WebGpuCommandEncoder>(command_encoder_rid)?;
+  let command_encoder = command_encoder_resource.1;
+  let query_set_resource = state
+    .resource_table
+    .get::<super::WebGpuQuerySet>(query_set)?;
+
+  gfx_ok!(command_encoder => instance.command_encoder_write_timestamp(
+    command_encoder,
+    query_set_resource.1,
+    query_index
+  ))
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_command_encoder_resolve_query_set(
+  state: &mut OpState,
+  #[smi] command_encoder_rid: ResourceId,
+  #[smi] query_set: ResourceId,
+  first_query: u32,
+  query_count: u32,
+  #[smi] destination: ResourceId,
+  #[number] destination_offset: u64,
+) -> Result<WebGpuResult, AnyError> {
+  let instance = state.borrow::<super::Instance>();
+  let command_encoder_resource = state
+    .resource_table
+    .get::<WebGpuCommandEncoder>(command_encoder_rid)?;
+  let command_encoder = command_encoder_resource.1;
+  let query_set_resource = state
+    .resource_table
+    .get::<super::WebGpuQuerySet>(query_set)?;
+  let destination_resource = state
+    .resource_table
+    .get::<super::buffer::WebGpuBuffer>(destination)?;
+
+  gfx_ok!(command_encoder => instance.command_encoder_resolve_query_set(
+    command_encoder,
+    query_set_resource.1,
+    first_query,
+    query_count,
+    destination_resource.1,
+    destination_offset
+  ))
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_command_encoder_finish(
+  state: &mut OpState,
+  #[smi] command_encoder_rid: ResourceId,
+  #[string] label: Cow<str>,
+) -> Result<WebGpuResult, AnyError> {
+  let command_encoder_resource = state
+    .resource_table
+    .take::<WebGpuCommandEncoder>(command_encoder_rid)?;
+  let command_encoder = command_encoder_resource.1;
+  let instance = state.borrow::<super::Instance>();
+
+  let descriptor = wgpu_types::CommandBufferDescriptor { label: Some(label) };
+
+  let (val, maybe_err) = gfx_select!(command_encoder => instance.command_encoder_finish(
+    command_encoder,
+    &descriptor
+  ));
+
+  let rid = state.resource_table.add(WebGpuCommandBuffer(
+    instance.clone(),
+    RefCell::new(Some(val)),
+  ));
+
+  Ok(WebGpuResult::rid_err(rid, maybe_err))
+}

ext/webgpu/compute_pass.rs

@@ -0,0 +1,225 @@
+// Copyright 2018-2023 the Deno authors. All rights reserved. MIT license.
+
+use deno_core::error::AnyError;
+use deno_core::op2;
+use deno_core::OpState;
+use deno_core::Resource;
+use deno_core::ResourceId;
+use std::borrow::Cow;
+use std::cell::RefCell;
+
+use super::error::WebGpuResult;
+
+pub(crate) struct WebGpuComputePass(
+  pub(crate) RefCell<wgpu_core::command::ComputePass>,
+);
+impl Resource for WebGpuComputePass {
+  fn name(&self) -> Cow<str> {
+    "webGPUComputePass".into()
+  }
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_compute_pass_set_pipeline(
+  state: &mut OpState,
+  #[smi] compute_pass_rid: ResourceId,
+  #[smi] pipeline: ResourceId,
+) -> Result<WebGpuResult, AnyError> {
+  let compute_pipeline_resource =
+    state
+      .resource_table
+      .get::<super::pipeline::WebGpuComputePipeline>(pipeline)?;
+  let compute_pass_resource = state
+    .resource_table
+    .get::<WebGpuComputePass>(compute_pass_rid)?;
+
+  wgpu_core::command::compute_ffi::wgpu_compute_pass_set_pipeline(
+    &mut compute_pass_resource.0.borrow_mut(),
+    compute_pipeline_resource.1,
+  );
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_compute_pass_dispatch_workgroups(
+  state: &mut OpState,
+  #[smi] compute_pass_rid: ResourceId,
+  x: u32,
+  y: u32,
+  z: u32,
+) -> Result<WebGpuResult, AnyError> {
+  let compute_pass_resource = state
+    .resource_table
+    .get::<WebGpuComputePass>(compute_pass_rid)?;
+
+  wgpu_core::command::compute_ffi::wgpu_compute_pass_dispatch_workgroups(
+    &mut compute_pass_resource.0.borrow_mut(),
+    x,
+    y,
+    z,
+  );
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_compute_pass_dispatch_workgroups_indirect(
+  state: &mut OpState,
+  #[smi] compute_pass_rid: ResourceId,
+  #[smi] indirect_buffer: ResourceId,
+  #[number] indirect_offset: u64,
+) -> Result<WebGpuResult, AnyError> {
+  let buffer_resource = state
+    .resource_table
+    .get::<super::buffer::WebGpuBuffer>(indirect_buffer)?;
+  let compute_pass_resource = state
+    .resource_table
+    .get::<WebGpuComputePass>(compute_pass_rid)?;
+
+  wgpu_core::command::compute_ffi::wgpu_compute_pass_dispatch_workgroups_indirect(
+        &mut compute_pass_resource.0.borrow_mut(),
+        buffer_resource.1,
+        indirect_offset,
+    );
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_compute_pass_end(
+  state: &mut OpState,
+  #[smi] command_encoder_rid: ResourceId,
+  #[smi] compute_pass_rid: ResourceId,
+) -> Result<WebGpuResult, AnyError> {
+  let command_encoder_resource = state
+    .resource_table
+    .get::<super::command_encoder::WebGpuCommandEncoder>(
+    command_encoder_rid,
+  )?;
+  let command_encoder = command_encoder_resource.1;
+  let compute_pass_resource = state
+    .resource_table
+    .take::<WebGpuComputePass>(compute_pass_rid)?;
+  let compute_pass = &compute_pass_resource.0.borrow();
+  let instance = state.borrow::<super::Instance>();
+
+  gfx_ok!(command_encoder => instance.command_encoder_run_compute_pass(
+    command_encoder,
+    compute_pass
+  ))
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_compute_pass_set_bind_group(
+  state: &mut OpState,
+  #[smi] compute_pass_rid: ResourceId,
+  index: u32,
+  #[smi] bind_group: ResourceId,
+  #[buffer] dynamic_offsets_data: &[u32],
+  #[number] dynamic_offsets_data_start: usize,
+  #[number] dynamic_offsets_data_length: usize,
+) -> Result<WebGpuResult, AnyError> {
+  let bind_group_resource =
+    state
+      .resource_table
+      .get::<super::binding::WebGpuBindGroup>(bind_group)?;
+  let compute_pass_resource = state
+    .resource_table
+    .get::<WebGpuComputePass>(compute_pass_rid)?;
+
+  let start = dynamic_offsets_data_start;
+  let len = dynamic_offsets_data_length;
+
+  // Assert that length and start are both in bounds
+  assert!(start <= dynamic_offsets_data.len());
+  assert!(len <= dynamic_offsets_data.len() - start);
+
+  let dynamic_offsets_data: &[u32] = &dynamic_offsets_data[start..start + len];
+
+  // SAFETY: the raw pointer and length are of the same slice, and that slice
+  // lives longer than the below function invocation.
+  unsafe {
+    wgpu_core::command::compute_ffi::wgpu_compute_pass_set_bind_group(
+      &mut compute_pass_resource.0.borrow_mut(),
+      index,
+      bind_group_resource.1,
+      dynamic_offsets_data.as_ptr(),
+      dynamic_offsets_data.len(),
+    );
+  }
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_compute_pass_push_debug_group(
+  state: &mut OpState,
+  #[smi] compute_pass_rid: ResourceId,
+  #[string] group_label: &str,
+) -> Result<WebGpuResult, AnyError> {
+  let compute_pass_resource = state
+    .resource_table
+    .get::<WebGpuComputePass>(compute_pass_rid)?;
+
+  let label = std::ffi::CString::new(group_label).unwrap();
+  // SAFETY: the string the raw pointer points to lives longer than the below
+  // function invocation.
+  unsafe {
+    wgpu_core::command::compute_ffi::wgpu_compute_pass_push_debug_group(
+      &mut compute_pass_resource.0.borrow_mut(),
+      label.as_ptr(),
+      0, // wgpu#975
+    );
+  }
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_compute_pass_pop_debug_group(
+  state: &mut OpState,
+  #[smi] compute_pass_rid: ResourceId,
+) -> Result<WebGpuResult, AnyError> {
+  let compute_pass_resource = state
+    .resource_table
+    .get::<WebGpuComputePass>(compute_pass_rid)?;
+
+  wgpu_core::command::compute_ffi::wgpu_compute_pass_pop_debug_group(
+    &mut compute_pass_resource.0.borrow_mut(),
+  );
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_compute_pass_insert_debug_marker(
+  state: &mut OpState,
+  #[smi] compute_pass_rid: ResourceId,
+  #[string] marker_label: &str,
+) -> Result<WebGpuResult, AnyError> {
+  let compute_pass_resource = state
+    .resource_table
+    .get::<WebGpuComputePass>(compute_pass_rid)?;
+
+  let label = std::ffi::CString::new(marker_label).unwrap();
+  // SAFETY: the string the raw pointer points to lives longer than the below
+  // function invocation.
+  unsafe {
+    wgpu_core::command::compute_ffi::wgpu_compute_pass_insert_debug_marker(
+      &mut compute_pass_resource.0.borrow_mut(),
+      label.as_ptr(),
+      0, // wgpu#975
+    );
+  }
+
+  Ok(WebGpuResult::empty())
+}

ext/webgpu/error.rs

@@ -0,0 +1,316 @@
+// Copyright 2018-2023 the Deno authors. All rights reserved. MIT license.
+use deno_core::error::AnyError;
+use deno_core::ResourceId;
+use serde::Serialize;
+use std::convert::From;
+use std::error::Error;
+use std::fmt;
+use wgpu_core::binding_model::CreateBindGroupError;
+use wgpu_core::binding_model::CreateBindGroupLayoutError;
+use wgpu_core::binding_model::CreatePipelineLayoutError;
+use wgpu_core::binding_model::GetBindGroupLayoutError;
+use wgpu_core::command::ClearError;
+use wgpu_core::command::CommandEncoderError;
+use wgpu_core::command::ComputePassError;
+use wgpu_core::command::CopyError;
+use wgpu_core::command::CreateRenderBundleError;
+use wgpu_core::command::QueryError;
+use wgpu_core::command::RenderBundleError;
+use wgpu_core::command::RenderPassError;
+use wgpu_core::device::queue::QueueSubmitError;
+use wgpu_core::device::queue::QueueWriteError;
+use wgpu_core::device::DeviceError;
+use wgpu_core::pipeline::CreateComputePipelineError;
+use wgpu_core::pipeline::CreateRenderPipelineError;
+use wgpu_core::pipeline::CreateShaderModuleError;
+#[cfg(feature = "surface")]
+use wgpu_core::present::ConfigureSurfaceError;
+use wgpu_core::resource::BufferAccessError;
+use wgpu_core::resource::CreateBufferError;
+use wgpu_core::resource::CreateQuerySetError;
+use wgpu_core::resource::CreateSamplerError;
+use wgpu_core::resource::CreateTextureError;
+use wgpu_core::resource::CreateTextureViewError;
+
+fn fmt_err(err: &(dyn Error + 'static)) -> String {
+  let mut output = err.to_string();
+
+  let mut e = err.source();
+  while let Some(source) = e {
+    output.push_str(&format!(": {source}"));
+    e = source.source();
+  }
+
+  output
+}
+
+#[derive(Serialize)]
+pub struct WebGpuResult {
+  pub rid: Option<ResourceId>,
+  pub err: Option<WebGpuError>,
+}
+
+impl WebGpuResult {
+  pub fn rid(rid: ResourceId) -> Self {
+    Self {
+      rid: Some(rid),
+      err: None,
+    }
+  }
+
+  pub fn rid_err<T: Into<WebGpuError>>(
+    rid: ResourceId,
+    err: Option<T>,
+  ) -> Self {
+    Self {
+      rid: Some(rid),
+      err: err.map(Into::into),
+    }
+  }
+
+  pub fn maybe_err<T: Into<WebGpuError>>(err: Option<T>) -> Self {
+    Self {
+      rid: None,
+      err: err.map(Into::into),
+    }
+  }
+
+  pub fn empty() -> Self {
+    Self {
+      rid: None,
+      err: None,
+    }
+  }
+}
+
+#[derive(Serialize)]
+#[serde(tag = "type", content = "value")]
+#[serde(rename_all = "kebab-case")]
+pub enum WebGpuError {
+  Lost,
+  OutOfMemory,
+  Validation(String),
+}
+
+impl From<CreateBufferError> for WebGpuError {
+  fn from(err: CreateBufferError) -> Self {
+    match err {
+      CreateBufferError::Device(err) => err.into(),
+      CreateBufferError::AccessError(err) => err.into(),
+      err => WebGpuError::Validation(fmt_err(&err)),
+    }
+  }
+}
+
+impl From<DeviceError> for WebGpuError {
+  fn from(err: DeviceError) -> Self {
+    match err {
+      DeviceError::Lost => WebGpuError::Lost,
+      DeviceError::OutOfMemory => WebGpuError::OutOfMemory,
+      DeviceError::ResourceCreationFailed
+      | DeviceError::Invalid
+      | DeviceError::WrongDevice => WebGpuError::Validation(fmt_err(&err)),
+    }
+  }
+}
+
+impl From<BufferAccessError> for WebGpuError {
+  fn from(err: BufferAccessError) -> Self {
+    match err {
+      BufferAccessError::Device(err) => err.into(),
+      err => WebGpuError::Validation(fmt_err(&err)),
+    }
+  }
+}
+
+impl From<CreateBindGroupLayoutError> for WebGpuError {
+  fn from(err: CreateBindGroupLayoutError) -> Self {
+    match err {
+      CreateBindGroupLayoutError::Device(err) => err.into(),
+      err => WebGpuError::Validation(fmt_err(&err)),
+    }
+  }
+}
+
+impl From<CreatePipelineLayoutError> for WebGpuError {
+  fn from(err: CreatePipelineLayoutError) -> Self {
+    match err {
+      CreatePipelineLayoutError::Device(err) => err.into(),
+      err => WebGpuError::Validation(fmt_err(&err)),
+    }
+  }
+}
+
+impl From<CreateBindGroupError> for WebGpuError {
+  fn from(err: CreateBindGroupError) -> Self {
+    match err {
+      CreateBindGroupError::Device(err) => err.into(),
+      err => WebGpuError::Validation(fmt_err(&err)),
+    }
+  }
+}
+
+impl From<RenderBundleError> for WebGpuError {
+  fn from(err: RenderBundleError) -> Self {
+    WebGpuError::Validation(fmt_err(&err))
+  }
+}
+
+impl From<CreateRenderBundleError> for WebGpuError {
+  fn from(err: CreateRenderBundleError) -> Self {
+    WebGpuError::Validation(fmt_err(&err))
+  }
+}
+
+impl From<CopyError> for WebGpuError {
+  fn from(err: CopyError) -> Self {
+    WebGpuError::Validation(fmt_err(&err))
+  }
+}
+
+impl From<CommandEncoderError> for WebGpuError {
+  fn from(err: CommandEncoderError) -> Self {
+    WebGpuError::Validation(fmt_err(&err))
+  }
+}
+
+impl From<QueryError> for WebGpuError {
+  fn from(err: QueryError) -> Self {
+    WebGpuError::Validation(fmt_err(&err))
+  }
+}
+
+impl From<ComputePassError> for WebGpuError {
+  fn from(err: ComputePassError) -> Self {
+    WebGpuError::Validation(fmt_err(&err))
+  }
+}
+
+impl From<CreateComputePipelineError> for WebGpuError {
+  fn from(err: CreateComputePipelineError) -> Self {
+    match err {
+      CreateComputePipelineError::Device(err) => err.into(),
+      err => WebGpuError::Validation(fmt_err(&err)),
+    }
+  }
+}
+
+impl From<GetBindGroupLayoutError> for WebGpuError {
+  fn from(err: GetBindGroupLayoutError) -> Self {
+    WebGpuError::Validation(fmt_err(&err))
+  }
+}
+
+impl From<CreateRenderPipelineError> for WebGpuError {
+  fn from(err: CreateRenderPipelineError) -> Self {
+    match err {
+      CreateRenderPipelineError::Device(err) => err.into(),
+      err => WebGpuError::Validation(fmt_err(&err)),
+    }
+  }
+}
+
+impl From<RenderPassError> for WebGpuError {
+  fn from(err: RenderPassError) -> Self {
+    WebGpuError::Validation(fmt_err(&err))
+  }
+}
+
+impl From<CreateSamplerError> for WebGpuError {
+  fn from(err: CreateSamplerError) -> Self {
+    match err {
+      CreateSamplerError::Device(err) => err.into(),
+      err => WebGpuError::Validation(fmt_err(&err)),
+    }
+  }
+}
+
+impl From<CreateShaderModuleError> for WebGpuError {
+  fn from(err: CreateShaderModuleError) -> Self {
+    match err {
+      CreateShaderModuleError::Device(err) => err.into(),
+      err => WebGpuError::Validation(fmt_err(&err)),
+    }
+  }
+}
+
+impl From<CreateTextureError> for WebGpuError {
+  fn from(err: CreateTextureError) -> Self {
+    match err {
+      CreateTextureError::Device(err) => err.into(),
+      err => WebGpuError::Validation(fmt_err(&err)),
+    }
+  }
+}
+
+impl From<CreateTextureViewError> for WebGpuError {
+  fn from(err: CreateTextureViewError) -> Self {
+    WebGpuError::Validation(fmt_err(&err))
+  }
+}
+
+impl From<CreateQuerySetError> for WebGpuError {
+  fn from(err: CreateQuerySetError) -> Self {
+    match err {
+      CreateQuerySetError::Device(err) => err.into(),
+      err => WebGpuError::Validation(fmt_err(&err)),
+    }
+  }
+}
+
+impl From<QueueSubmitError> for WebGpuError {
+  fn from(err: QueueSubmitError) -> Self {
+    match err {
+      QueueSubmitError::Queue(err) => err.into(),
+      err => WebGpuError::Validation(fmt_err(&err)),
+    }
+  }
+}
+
+impl From<QueueWriteError> for WebGpuError {
+  fn from(err: QueueWriteError) -> Self {
+    match err {
+      QueueWriteError::Queue(err) => err.into(),
+      err => WebGpuError::Validation(fmt_err(&err)),
+    }
+  }
+}
+
+impl From<ClearError> for WebGpuError {
+  fn from(err: ClearError) -> Self {
+    WebGpuError::Validation(fmt_err(&err))
+  }
+}
+
+#[cfg(feature = "surface")]
+impl From<ConfigureSurfaceError> for WebGpuError {
+  fn from(err: ConfigureSurfaceError) -> Self {
+    WebGpuError::Validation(fmt_err(&err))
+  }
+}
+
+#[derive(Debug)]
+pub struct DomExceptionOperationError {
+  pub msg: String,
+}
+
+impl DomExceptionOperationError {
+  pub fn new(msg: &str) -> Self {
+    DomExceptionOperationError {
+      msg: msg.to_string(),
+    }
+  }
+}
+
+impl fmt::Display for DomExceptionOperationError {
+  fn fmt(&self, f: &mut fmt::Formatter) -> fmt::Result {
+    f.pad(&self.msg)
+  }
+}
+
+impl std::error::Error for DomExceptionOperationError {}
+
+pub fn get_error_class_name(e: &AnyError) -> Option<&'static str> {
+  e.downcast_ref::<DomExceptionOperationError>()
+    .map(|_| "DOMExceptionOperationError")
+}

ext/webgpu/lib.rs

@@ -0,0 +1,768 @@
+// Copyright 2018-2023 the Deno authors. All rights reserved. MIT license.
+#![cfg(not(target_arch = "wasm32"))]
+#![warn(unsafe_op_in_unsafe_fn)]
+
+use deno_core::error::AnyError;
+use deno_core::op2;
+use deno_core::OpState;
+use deno_core::Resource;
+use deno_core::ResourceId;
+use serde::Deserialize;
+use serde::Serialize;
+use std::borrow::Cow;
+use std::cell::RefCell;
+use std::collections::HashSet;
+use std::rc::Rc;
+pub use wgpu_core;
+pub use wgpu_types;
+
+use error::DomExceptionOperationError;
+use error::WebGpuResult;
+
+pub const UNSTABLE_FEATURE_NAME: &str = "webgpu";
+
+#[macro_use]
+mod macros {
+  macro_rules! gfx_select {
+    ($id:expr => $global:ident.$method:ident( $($param:expr),* )) => {
+      match $id.backend() {
+        #[cfg(any(
+            all(not(target_arch = "wasm32"), not(target_os = "ios"), not(target_os = "macos")),
+            feature = "vulkan-portability"
+        ))]
+        wgpu_types::Backend::Vulkan => $global.$method::<wgpu_core::api::Vulkan>( $($param),* ),
+        #[cfg(all(not(target_arch = "wasm32"), any(target_os = "ios", target_os = "macos")))]
+        wgpu_types::Backend::Metal => $global.$method::<wgpu_core::api::Metal>( $($param),* ),
+        #[cfg(all(not(target_arch = "wasm32"), windows))]
+        wgpu_types::Backend::Dx12 => $global.$method::<wgpu_core::api::Dx12>( $($param),* ),
+        #[cfg(all(not(target_arch = "wasm32"), windows))]
+        wgpu_types::Backend::Dx11 => $global.$method::<wgpu_core::api::Dx11>( $($param),* ),
+        #[cfg(any(
+            all(unix, not(target_os = "macos"), not(target_os = "ios")),
+            feature = "angle",
+            target_arch = "wasm32"
+        ))]
+        wgpu_types::Backend::Gl => $global.$method::<wgpu_core::api::Gles>( $($param),+ ),
+        other => panic!("Unexpected backend {:?}", other),
+      }
+    };
+  }
+
+  macro_rules! gfx_put {
+    ($id:expr => $global:ident.$method:ident( $($param:expr),* ) => $state:expr, $rc:expr) => {{
+      let (val, maybe_err) = gfx_select!($id => $global.$method($($param),*));
+      let rid = $state.resource_table.add($rc($global.clone(), val));
+      Ok(WebGpuResult::rid_err(rid, maybe_err))
+    }};
+  }
+
+  macro_rules! gfx_ok {
+    ($id:expr => $global:ident.$method:ident( $($param:expr),* )) => {{
+      let maybe_err = gfx_select!($id => $global.$method($($param),*)).err();
+      Ok(WebGpuResult::maybe_err(maybe_err))
+    }};
+  }
+}
+
+pub mod binding;
+pub mod buffer;
+pub mod bundle;
+pub mod command_encoder;
+pub mod compute_pass;
+pub mod error;
+pub mod pipeline;
+pub mod queue;
+pub mod render_pass;
+pub mod sampler;
+pub mod shader;
+#[cfg(feature = "surface")]
+pub mod surface;
+pub mod texture;
+
+pub type Instance = std::sync::Arc<
+  wgpu_core::global::Global<wgpu_core::identity::IdentityManagerFactory>,
+>;
+
+struct WebGpuAdapter(Instance, wgpu_core::id::AdapterId);
+impl Resource for WebGpuAdapter {
+  fn name(&self) -> Cow<str> {
+    "webGPUAdapter".into()
+  }
+
+  fn close(self: Rc<Self>) {
+    let instance = &self.0;
+    gfx_select!(self.1 => instance.adapter_drop(self.1));
+  }
+}
+
+struct WebGpuDevice(Instance, wgpu_core::id::DeviceId);
+impl Resource for WebGpuDevice {
+  fn name(&self) -> Cow<str> {
+    "webGPUDevice".into()
+  }
+
+  fn close(self: Rc<Self>) {
+    let instance = &self.0;
+    gfx_select!(self.1 => instance.device_drop(self.1));
+  }
+}
+
+struct WebGpuQuerySet(Instance, wgpu_core::id::QuerySetId);
+impl Resource for WebGpuQuerySet {
+  fn name(&self) -> Cow<str> {
+    "webGPUQuerySet".into()
+  }
+
+  fn close(self: Rc<Self>) {
+    let instance = &self.0;
+    gfx_select!(self.1 => instance.query_set_drop(self.1));
+  }
+}
+
+deno_core::extension!(
+  deno_webgpu,
+  deps = [deno_webidl, deno_web],
+  ops = [
+    // Request device/adapter
+    op_webgpu_request_adapter,
+    op_webgpu_request_device,
+    op_webgpu_request_adapter_info,
+    // Query Set
+    op_webgpu_create_query_set,
+    // buffer
+    buffer::op_webgpu_create_buffer,
+    buffer::op_webgpu_buffer_get_mapped_range,
+    buffer::op_webgpu_buffer_unmap,
+    // buffer async
+    buffer::op_webgpu_buffer_get_map_async,
+    // remaining sync ops
+
+    // texture
+    texture::op_webgpu_create_texture,
+    texture::op_webgpu_create_texture_view,
+    // sampler
+    sampler::op_webgpu_create_sampler,
+    // binding
+    binding::op_webgpu_create_bind_group_layout,
+    binding::op_webgpu_create_pipeline_layout,
+    binding::op_webgpu_create_bind_group,
+    // pipeline
+    pipeline::op_webgpu_create_compute_pipeline,
+    pipeline::op_webgpu_compute_pipeline_get_bind_group_layout,
+    pipeline::op_webgpu_create_render_pipeline,
+    pipeline::op_webgpu_render_pipeline_get_bind_group_layout,
+    // command_encoder
+    command_encoder::op_webgpu_create_command_encoder,
+    command_encoder::op_webgpu_command_encoder_begin_render_pass,
+    command_encoder::op_webgpu_command_encoder_begin_compute_pass,
+    command_encoder::op_webgpu_command_encoder_copy_buffer_to_buffer,
+    command_encoder::op_webgpu_command_encoder_copy_buffer_to_texture,
+    command_encoder::op_webgpu_command_encoder_copy_texture_to_buffer,
+    command_encoder::op_webgpu_command_encoder_copy_texture_to_texture,
+    command_encoder::op_webgpu_command_encoder_clear_buffer,
+    command_encoder::op_webgpu_command_encoder_push_debug_group,
+    command_encoder::op_webgpu_command_encoder_pop_debug_group,
+    command_encoder::op_webgpu_command_encoder_insert_debug_marker,
+    command_encoder::op_webgpu_command_encoder_write_timestamp,
+    command_encoder::op_webgpu_command_encoder_resolve_query_set,
+    command_encoder::op_webgpu_command_encoder_finish,
+    render_pass::op_webgpu_render_pass_set_viewport,
+    render_pass::op_webgpu_render_pass_set_scissor_rect,
+    render_pass::op_webgpu_render_pass_set_blend_constant,
+    render_pass::op_webgpu_render_pass_set_stencil_reference,
+    render_pass::op_webgpu_render_pass_begin_occlusion_query,
+    render_pass::op_webgpu_render_pass_end_occlusion_query,
+    render_pass::op_webgpu_render_pass_execute_bundles,
+    render_pass::op_webgpu_render_pass_end,
+    render_pass::op_webgpu_render_pass_set_bind_group,
+    render_pass::op_webgpu_render_pass_push_debug_group,
+    render_pass::op_webgpu_render_pass_pop_debug_group,
+    render_pass::op_webgpu_render_pass_insert_debug_marker,
+    render_pass::op_webgpu_render_pass_set_pipeline,
+    render_pass::op_webgpu_render_pass_set_index_buffer,
+    render_pass::op_webgpu_render_pass_set_vertex_buffer,
+    render_pass::op_webgpu_render_pass_draw,
+    render_pass::op_webgpu_render_pass_draw_indexed,
+    render_pass::op_webgpu_render_pass_draw_indirect,
+    render_pass::op_webgpu_render_pass_draw_indexed_indirect,
+    compute_pass::op_webgpu_compute_pass_set_pipeline,
+    compute_pass::op_webgpu_compute_pass_dispatch_workgroups,
+    compute_pass::op_webgpu_compute_pass_dispatch_workgroups_indirect,
+    compute_pass::op_webgpu_compute_pass_end,
+    compute_pass::op_webgpu_compute_pass_set_bind_group,
+    compute_pass::op_webgpu_compute_pass_push_debug_group,
+    compute_pass::op_webgpu_compute_pass_pop_debug_group,
+    compute_pass::op_webgpu_compute_pass_insert_debug_marker,
+    // bundle
+    bundle::op_webgpu_create_render_bundle_encoder,
+    bundle::op_webgpu_render_bundle_encoder_finish,
+    bundle::op_webgpu_render_bundle_encoder_set_bind_group,
+    bundle::op_webgpu_render_bundle_encoder_push_debug_group,
+    bundle::op_webgpu_render_bundle_encoder_pop_debug_group,
+    bundle::op_webgpu_render_bundle_encoder_insert_debug_marker,
+    bundle::op_webgpu_render_bundle_encoder_set_pipeline,
+    bundle::op_webgpu_render_bundle_encoder_set_index_buffer,
+    bundle::op_webgpu_render_bundle_encoder_set_vertex_buffer,
+    bundle::op_webgpu_render_bundle_encoder_draw,
+    bundle::op_webgpu_render_bundle_encoder_draw_indexed,
+    bundle::op_webgpu_render_bundle_encoder_draw_indirect,
+    // queue
+    queue::op_webgpu_queue_submit,
+    queue::op_webgpu_write_buffer,
+    queue::op_webgpu_write_texture,
+    // shader
+    shader::op_webgpu_create_shader_module,
+  ],
+  lazy_loaded_esm = ["01_webgpu.js"],
+);
+
+fn deserialize_features(features: &wgpu_types::Features) -> Vec<&'static str> {
+  let mut return_features: Vec<&'static str> = vec![];
+
+  // api
+  if features.contains(wgpu_types::Features::DEPTH_CLIP_CONTROL) {
+    return_features.push("depth-clip-control");
+  }
+  if features.contains(wgpu_types::Features::TIMESTAMP_QUERY) {
+    return_features.push("timestamp-query");
+  }
+  if features.contains(wgpu_types::Features::INDIRECT_FIRST_INSTANCE) {
+    return_features.push("indirect-first-instance");
+  }
+  // shader
+  if features.contains(wgpu_types::Features::SHADER_F16) {
+    return_features.push("shader-f16");
+  }
+  // texture formats
+  if features.contains(wgpu_types::Features::DEPTH32FLOAT_STENCIL8) {
+    return_features.push("depth32float-stencil8");
+  }
+  if features.contains(wgpu_types::Features::TEXTURE_COMPRESSION_BC) {
+    return_features.push("texture-compression-bc");
+  }
+  if features.contains(wgpu_types::Features::TEXTURE_COMPRESSION_ETC2) {
+    return_features.push("texture-compression-etc2");
+  }
+  if features.contains(wgpu_types::Features::TEXTURE_COMPRESSION_ASTC) {
+    return_features.push("texture-compression-astc");
+  }
+  if features.contains(wgpu_types::Features::RG11B10UFLOAT_RENDERABLE) {
+    return_features.push("rg11b10ufloat-renderable");
+  }
+  if features.contains(wgpu_types::Features::BGRA8UNORM_STORAGE) {
+    return_features.push("bgra8unorm-storage");
+  }
+
+  // extended from spec
+
+  // texture formats
+  if features.contains(wgpu_types::Features::TEXTURE_FORMAT_16BIT_NORM) {
+    return_features.push("texture-format-16-bit-norm");
+  }
+  if features.contains(wgpu_types::Features::TEXTURE_COMPRESSION_ASTC_HDR) {
+    return_features.push("texture-compression-astc-hdr");
+  }
+  if features
+    .contains(wgpu_types::Features::TEXTURE_ADAPTER_SPECIFIC_FORMAT_FEATURES)
+  {
+    return_features.push("texture-adapter-specific-format-features");
+  }
+  // api
+  if features.contains(wgpu_types::Features::PIPELINE_STATISTICS_QUERY) {
+    return_features.push("pipeline-statistics-query");
+  }
+  if features.contains(wgpu_types::Features::TIMESTAMP_QUERY_INSIDE_PASSES) {
+    return_features.push("timestamp-query-inside-passes");
+  }
+  if features.contains(wgpu_types::Features::MAPPABLE_PRIMARY_BUFFERS) {
+    return_features.push("mappable-primary-buffers");
+  }
+  if features.contains(wgpu_types::Features::TEXTURE_BINDING_ARRAY) {
+    return_features.push("texture-binding-array");
+  }
+  if features.contains(wgpu_types::Features::BUFFER_BINDING_ARRAY) {
+    return_features.push("buffer-binding-array");
+  }
+  if features.contains(wgpu_types::Features::STORAGE_RESOURCE_BINDING_ARRAY) {
+    return_features.push("storage-resource-binding-array");
+  }
+  if features.contains(
+        wgpu_types::Features::SAMPLED_TEXTURE_AND_STORAGE_BUFFER_ARRAY_NON_UNIFORM_INDEXING,
+    ) {
+        return_features.push("sampled-texture-and-storage-buffer-array-non-uniform-indexing");
+    }
+  if features.contains(
+        wgpu_types::Features::UNIFORM_BUFFER_AND_STORAGE_TEXTURE_ARRAY_NON_UNIFORM_INDEXING,
+    ) {
+        return_features.push("uniform-buffer-and-storage-texture-array-non-uniform-indexing");
+    }
+  if features.contains(wgpu_types::Features::PARTIALLY_BOUND_BINDING_ARRAY) {
+    return_features.push("partially-bound-binding-array");
+  }
+  if features.contains(wgpu_types::Features::MULTI_DRAW_INDIRECT) {
+    return_features.push("multi-draw-indirect");
+  }
+  if features.contains(wgpu_types::Features::MULTI_DRAW_INDIRECT_COUNT) {
+    return_features.push("multi-draw-indirect-count");
+  }
+  if features.contains(wgpu_types::Features::PUSH_CONSTANTS) {
+    return_features.push("push-constants");
+  }
+  if features.contains(wgpu_types::Features::ADDRESS_MODE_CLAMP_TO_ZERO) {
+    return_features.push("address-mode-clamp-to-zero");
+  }
+  if features.contains(wgpu_types::Features::ADDRESS_MODE_CLAMP_TO_BORDER) {
+    return_features.push("address-mode-clamp-to-border");
+  }
+  if features.contains(wgpu_types::Features::POLYGON_MODE_LINE) {
+    return_features.push("polygon-mode-line");
+  }
+  if features.contains(wgpu_types::Features::POLYGON_MODE_POINT) {
+    return_features.push("polygon-mode-point");
+  }
+  if features.contains(wgpu_types::Features::CONSERVATIVE_RASTERIZATION) {
+    return_features.push("conservative-rasterization");
+  }
+  if features.contains(wgpu_types::Features::VERTEX_WRITABLE_STORAGE) {
+    return_features.push("vertex-writable-storage");
+  }
+  if features.contains(wgpu_types::Features::CLEAR_TEXTURE) {
+    return_features.push("clear-texture");
+  }
+  if features.contains(wgpu_types::Features::SPIRV_SHADER_PASSTHROUGH) {
+    return_features.push("spirv-shader-passthrough");
+  }
+  if features.contains(wgpu_types::Features::MULTIVIEW) {
+    return_features.push("multiview");
+  }
+  if features.contains(wgpu_types::Features::VERTEX_ATTRIBUTE_64BIT) {
+    return_features.push("vertex-attribute-64-bit");
+  }
+  // shader
+  if features.contains(wgpu_types::Features::SHADER_F64) {
+    return_features.push("shader-f64");
+  }
+  if features.contains(wgpu_types::Features::SHADER_I16) {
+    return_features.push("shader-i16");
+  }
+  if features.contains(wgpu_types::Features::SHADER_PRIMITIVE_INDEX) {
+    return_features.push("shader-primitive-index");
+  }
+  if features.contains(wgpu_types::Features::SHADER_EARLY_DEPTH_TEST) {
+    return_features.push("shader-early-depth-test");
+  }
+  if features.contains(wgpu_types::Features::SHADER_UNUSED_VERTEX_OUTPUT) {
+    return_features.push("shader-unused-vertex-output");
+  }
+
+  return_features
+}
+
+#[derive(Serialize)]
+#[serde(untagged)]
+pub enum GpuAdapterDeviceOrErr {
+  Error { err: String },
+  Features(GpuAdapterDevice),
+}
+
+#[derive(Serialize)]
+#[serde(rename_all = "camelCase")]
+pub struct GpuAdapterDevice {
+  rid: ResourceId,
+  limits: wgpu_types::Limits,
+  features: Vec<&'static str>,
+  is_software: bool,
+}
+
+#[op2(async)]
+#[serde]
+pub async fn op_webgpu_request_adapter(
+  state: Rc<RefCell<OpState>>,
+  #[serde] power_preference: Option<wgpu_types::PowerPreference>,
+  force_fallback_adapter: bool,
+) -> Result<GpuAdapterDeviceOrErr, AnyError> {
+  let mut state = state.borrow_mut();
+
+  // TODO(bartlomieju): replace with `state.feature_checker.check_or_exit`
+  // once we phase out `check_or_exit_with_legacy_fallback`
+  state.feature_checker.check_or_exit_with_legacy_fallback(
+    UNSTABLE_FEATURE_NAME,
+    "navigator.gpu.requestAdapter",
+  );
+
+  let backends = std::env::var("DENO_WEBGPU_BACKEND").map_or_else(
+    |_| wgpu_types::Backends::all(),
+    |s| wgpu_core::instance::parse_backends_from_comma_list(&s),
+  );
+  let instance = if let Some(instance) = state.try_borrow::<Instance>() {
+    instance
+  } else {
+    state.put(std::sync::Arc::new(wgpu_core::global::Global::new(
+      "webgpu",
+      wgpu_core::identity::IdentityManagerFactory,
+      wgpu_types::InstanceDescriptor {
+        backends,
+        flags: wgpu_types::InstanceFlags::from_build_config(),
+        dx12_shader_compiler: wgpu_types::Dx12Compiler::Fxc,
+        gles_minor_version: wgpu_types::Gles3MinorVersion::default(),
+      },
+    )));
+    state.borrow::<Instance>()
+  };
+
+  let descriptor = wgpu_core::instance::RequestAdapterOptions {
+    power_preference: power_preference.unwrap_or_default(),
+    force_fallback_adapter,
+    compatible_surface: None, // windowless
+  };
+  let res = instance.request_adapter(
+    &descriptor,
+    wgpu_core::instance::AdapterInputs::Mask(backends, |_| ()),
+  );
+
+  let adapter = match res {
+    Ok(adapter) => adapter,
+    Err(err) => {
+      return Ok(GpuAdapterDeviceOrErr::Error {
+        err: err.to_string(),
+      })
+    }
+  };
+  let adapter_features =
+    gfx_select!(adapter => instance.adapter_features(adapter))?;
+  let features = deserialize_features(&adapter_features);
+  let adapter_limits =
+    gfx_select!(adapter => instance.adapter_limits(adapter))?;
+
+  let instance = instance.clone();
+
+  let rid = state.resource_table.add(WebGpuAdapter(instance, adapter));
+
+  Ok(GpuAdapterDeviceOrErr::Features(GpuAdapterDevice {
+    rid,
+    features,
+    limits: adapter_limits,
+    is_software: false,
+  }))
+}
+
+#[derive(Deserialize)]
+pub struct GpuRequiredFeatures(HashSet<String>);
+
+impl From<GpuRequiredFeatures> for wgpu_types::Features {
+  fn from(required_features: GpuRequiredFeatures) -> wgpu_types::Features {
+    let mut features: wgpu_types::Features = wgpu_types::Features::empty();
+    // api
+    features.set(
+      wgpu_types::Features::DEPTH_CLIP_CONTROL,
+      required_features.0.contains("depth-clip-control"),
+    );
+    features.set(
+      wgpu_types::Features::TIMESTAMP_QUERY,
+      required_features.0.contains("timestamp-query"),
+    );
+    features.set(
+      wgpu_types::Features::INDIRECT_FIRST_INSTANCE,
+      required_features.0.contains("indirect-first-instance"),
+    );
+    // shader
+    features.set(
+      wgpu_types::Features::SHADER_F16,
+      required_features.0.contains("shader-f16"),
+    );
+    // texture formats
+    features.set(
+      wgpu_types::Features::DEPTH32FLOAT_STENCIL8,
+      required_features.0.contains("depth32float-stencil8"),
+    );
+    features.set(
+      wgpu_types::Features::TEXTURE_COMPRESSION_BC,
+      required_features.0.contains("texture-compression-bc"),
+    );
+    features.set(
+      wgpu_types::Features::TEXTURE_COMPRESSION_ETC2,
+      required_features.0.contains("texture-compression-etc2"),
+    );
+    features.set(
+      wgpu_types::Features::TEXTURE_COMPRESSION_ASTC,
+      required_features.0.contains("texture-compression-astc"),
+    );
+    features.set(
+      wgpu_types::Features::RG11B10UFLOAT_RENDERABLE,
+      required_features.0.contains("rg11b10ufloat-renderable"),
+    );
+    features.set(
+      wgpu_types::Features::BGRA8UNORM_STORAGE,
+      required_features.0.contains("bgra8unorm-storage"),
+    );
+
+    // extended from spec
+
+    // texture formats
+    features.set(
+      wgpu_types::Features::TEXTURE_FORMAT_16BIT_NORM,
+      required_features.0.contains("texture-format-16-bit-norm"),
+    );
+    features.set(
+      wgpu_types::Features::TEXTURE_COMPRESSION_ASTC_HDR,
+      required_features.0.contains("texture-compression-astc-hdr"),
+    );
+    features.set(
+      wgpu_types::Features::TEXTURE_ADAPTER_SPECIFIC_FORMAT_FEATURES,
+      required_features
+        .0
+        .contains("texture-adapter-specific-format-features"),
+    );
+    // api
+    features.set(
+      wgpu_types::Features::PIPELINE_STATISTICS_QUERY,
+      required_features.0.contains("pipeline-statistics-query"),
+    );
+    features.set(
+      wgpu_types::Features::TIMESTAMP_QUERY_INSIDE_PASSES,
+      required_features
+        .0
+        .contains("timestamp-query-inside-passes"),
+    );
+    features.set(
+      wgpu_types::Features::MAPPABLE_PRIMARY_BUFFERS,
+      required_features.0.contains("mappable-primary-buffers"),
+    );
+    features.set(
+      wgpu_types::Features::TEXTURE_BINDING_ARRAY,
+      required_features.0.contains("texture-binding-array"),
+    );
+    features.set(
+      wgpu_types::Features::BUFFER_BINDING_ARRAY,
+      required_features.0.contains("buffer-binding-array"),
+    );
+    features.set(
+      wgpu_types::Features::STORAGE_RESOURCE_BINDING_ARRAY,
+      required_features
+        .0
+        .contains("storage-resource-binding-array"),
+    );
+    features.set(
+            wgpu_types::Features::SAMPLED_TEXTURE_AND_STORAGE_BUFFER_ARRAY_NON_UNIFORM_INDEXING,
+            required_features
+                .0
+                .contains("sampled-texture-and-storage-buffer-array-non-uniform-indexing"),
+        );
+    features.set(
+            wgpu_types::Features::UNIFORM_BUFFER_AND_STORAGE_TEXTURE_ARRAY_NON_UNIFORM_INDEXING,
+            required_features
+                .0
+                .contains("uniform-buffer-and-storage-texture-array-non-uniform-indexing"),
+        );
+    features.set(
+      wgpu_types::Features::PARTIALLY_BOUND_BINDING_ARRAY,
+      required_features
+        .0
+        .contains("partially-bound-binding-array"),
+    );
+    features.set(
+      wgpu_types::Features::MULTI_DRAW_INDIRECT,
+      required_features.0.contains("multi-draw-indirect"),
+    );
+    features.set(
+      wgpu_types::Features::MULTI_DRAW_INDIRECT_COUNT,
+      required_features.0.contains("multi-draw-indirect-count"),
+    );
+    features.set(
+      wgpu_types::Features::PUSH_CONSTANTS,
+      required_features.0.contains("push-constants"),
+    );
+    features.set(
+      wgpu_types::Features::ADDRESS_MODE_CLAMP_TO_ZERO,
+      required_features.0.contains("address-mode-clamp-to-zero"),
+    );
+    features.set(
+      wgpu_types::Features::ADDRESS_MODE_CLAMP_TO_BORDER,
+      required_features.0.contains("address-mode-clamp-to-border"),
+    );
+    features.set(
+      wgpu_types::Features::POLYGON_MODE_LINE,
+      required_features.0.contains("polygon-mode-line"),
+    );
+    features.set(
+      wgpu_types::Features::POLYGON_MODE_POINT,
+      required_features.0.contains("polygon-mode-point"),
+    );
+    features.set(
+      wgpu_types::Features::CONSERVATIVE_RASTERIZATION,
+      required_features.0.contains("conservative-rasterization"),
+    );
+    features.set(
+      wgpu_types::Features::VERTEX_WRITABLE_STORAGE,
+      required_features.0.contains("vertex-writable-storage"),
+    );
+    features.set(
+      wgpu_types::Features::CLEAR_TEXTURE,
+      required_features.0.contains("clear-texture"),
+    );
+    features.set(
+      wgpu_types::Features::SPIRV_SHADER_PASSTHROUGH,
+      required_features.0.contains("spirv-shader-passthrough"),
+    );
+    features.set(
+      wgpu_types::Features::MULTIVIEW,
+      required_features.0.contains("multiview"),
+    );
+    features.set(
+      wgpu_types::Features::VERTEX_ATTRIBUTE_64BIT,
+      required_features.0.contains("vertex-attribute-64-bit"),
+    );
+    // shader
+    features.set(
+      wgpu_types::Features::SHADER_F64,
+      required_features.0.contains("shader-f64"),
+    );
+    features.set(
+      wgpu_types::Features::SHADER_I16,
+      required_features.0.contains("shader-i16"),
+    );
+    features.set(
+      wgpu_types::Features::SHADER_PRIMITIVE_INDEX,
+      required_features.0.contains("shader-primitive-index"),
+    );
+    features.set(
+      wgpu_types::Features::SHADER_EARLY_DEPTH_TEST,
+      required_features.0.contains("shader-early-depth-test"),
+    );
+    features.set(
+      wgpu_types::Features::SHADER_UNUSED_VERTEX_OUTPUT,
+      required_features.0.contains("shader-unused-vertex-output"),
+    );
+
+    features
+  }
+}
+
+#[op2(async)]
+#[serde]
+pub async fn op_webgpu_request_device(
+  state: Rc<RefCell<OpState>>,
+  #[smi] adapter_rid: ResourceId,
+  #[string] label: String,
+  #[serde] required_features: GpuRequiredFeatures,
+  #[serde] required_limits: Option<wgpu_types::Limits>,
+) -> Result<GpuAdapterDevice, AnyError> {
+  let mut state = state.borrow_mut();
+  let adapter_resource =
+    state.resource_table.get::<WebGpuAdapter>(adapter_rid)?;
+  let adapter = adapter_resource.1;
+  let instance = state.borrow::<Instance>();
+
+  let descriptor = wgpu_types::DeviceDescriptor {
+    label: Some(Cow::Owned(label)),
+    features: required_features.into(),
+    limits: required_limits.unwrap_or_default(),
+  };
+
+  let (device, maybe_err) = gfx_select!(adapter => instance.adapter_request_device(
+    adapter,
+    &descriptor,
+    std::env::var("DENO_WEBGPU_TRACE").ok().as_ref().map(std::path::Path::new),
+    ()
+  ));
+  if let Some(err) = maybe_err {
+    return Err(DomExceptionOperationError::new(&err.to_string()).into());
+  }
+
+  let device_features =
+    gfx_select!(device => instance.device_features(device))?;
+  let features = deserialize_features(&device_features);
+  let limits = gfx_select!(device => instance.device_limits(device))?;
+
+  let instance = instance.clone();
+  let rid = state.resource_table.add(WebGpuDevice(instance, device));
+
+  Ok(GpuAdapterDevice {
+    rid,
+    features,
+    limits,
+    // TODO(lucacasonato): report correctly from wgpu
+    is_software: false,
+  })
+}
+
+#[derive(Serialize)]
+#[serde(rename_all = "camelCase")]
+pub struct GPUAdapterInfo {
+  vendor: String,
+  architecture: String,
+  device: String,
+  description: String,
+}
+
+#[op2(async)]
+#[serde]
+pub async fn op_webgpu_request_adapter_info(
+  state: Rc<RefCell<OpState>>,
+  #[smi] adapter_rid: ResourceId,
+) -> Result<GPUAdapterInfo, AnyError> {
+  let state = state.borrow_mut();
+  let adapter_resource =
+    state.resource_table.get::<WebGpuAdapter>(adapter_rid)?;
+  let adapter = adapter_resource.1;
+  let instance = state.borrow::<Instance>();
+
+  let info = gfx_select!(adapter => instance.adapter_get_info(adapter))?;
+
+  Ok(GPUAdapterInfo {
+    vendor: info.vendor.to_string(),
+    architecture: String::new(), // TODO(#2170)
+    device: info.device.to_string(),
+    description: info.name,
+  })
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "camelCase")]
+pub struct CreateQuerySetArgs {
+  device_rid: ResourceId,
+  label: String,
+  #[serde(flatten)]
+  r#type: GpuQueryType,
+  count: u32,
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "kebab-case", tag = "type")]
+enum GpuQueryType {
+  Occlusion,
+  Timestamp,
+}
+
+impl From<GpuQueryType> for wgpu_types::QueryType {
+  fn from(query_type: GpuQueryType) -> Self {
+    match query_type {
+      GpuQueryType::Occlusion => wgpu_types::QueryType::Occlusion,
+      GpuQueryType::Timestamp => wgpu_types::QueryType::Timestamp,
+    }
+  }
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_create_query_set(
+  state: &mut OpState,
+  #[serde] args: CreateQuerySetArgs,
+) -> Result<WebGpuResult, AnyError> {
+  let device_resource =
+    state.resource_table.get::<WebGpuDevice>(args.device_rid)?;
+  let device = device_resource.1;
+  let instance = state.borrow::<Instance>();
+
+  let descriptor = wgpu_types::QuerySetDescriptor {
+    label: Some(Cow::Owned(args.label)),
+    ty: args.r#type.into(),
+    count: args.count,
+  };
+
+  gfx_put!(device => instance.device_create_query_set(
+    device,
+    &descriptor,
+    ()
+  ) => state, WebGpuQuerySet)
+}

ext/webgpu/pipeline.rs

@@ -0,0 +1,453 @@
+// Copyright 2018-2023 the Deno authors. All rights reserved. MIT license.
+
+use deno_core::error::AnyError;
+use deno_core::op2;
+use deno_core::OpState;
+use deno_core::Resource;
+use deno_core::ResourceId;
+use serde::Deserialize;
+use serde::Serialize;
+use std::borrow::Cow;
+use std::rc::Rc;
+
+use super::error::WebGpuError;
+use super::error::WebGpuResult;
+
+const MAX_BIND_GROUPS: usize = 8;
+
+pub(crate) struct WebGpuPipelineLayout(
+  pub(crate) crate::Instance,
+  pub(crate) wgpu_core::id::PipelineLayoutId,
+);
+impl Resource for WebGpuPipelineLayout {
+  fn name(&self) -> Cow<str> {
+    "webGPUPipelineLayout".into()
+  }
+
+  fn close(self: Rc<Self>) {
+    let instance = &self.0;
+    gfx_select!(self.1 => instance.pipeline_layout_drop(self.1));
+  }
+}
+
+pub(crate) struct WebGpuComputePipeline(
+  pub(crate) crate::Instance,
+  pub(crate) wgpu_core::id::ComputePipelineId,
+);
+impl Resource for WebGpuComputePipeline {
+  fn name(&self) -> Cow<str> {
+    "webGPUComputePipeline".into()
+  }
+
+  fn close(self: Rc<Self>) {
+    let instance = &self.0;
+    gfx_select!(self.1 => instance.compute_pipeline_drop(self.1));
+  }
+}
+
+pub(crate) struct WebGpuRenderPipeline(
+  pub(crate) crate::Instance,
+  pub(crate) wgpu_core::id::RenderPipelineId,
+);
+impl Resource for WebGpuRenderPipeline {
+  fn name(&self) -> Cow<str> {
+    "webGPURenderPipeline".into()
+  }
+
+  fn close(self: Rc<Self>) {
+    let instance = &self.0;
+    gfx_select!(self.1 => instance.render_pipeline_drop(self.1));
+  }
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "camelCase")]
+pub enum GPUAutoLayoutMode {
+  Auto,
+}
+
+#[derive(Deserialize)]
+#[serde(untagged)]
+pub enum GPUPipelineLayoutOrGPUAutoLayoutMode {
+  Layout(ResourceId),
+  Auto(GPUAutoLayoutMode),
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "camelCase")]
+pub struct GpuProgrammableStage {
+  module: ResourceId,
+  entry_point: String,
+  // constants: HashMap<String, GPUPipelineConstantValue>
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_create_compute_pipeline(
+  state: &mut OpState,
+  #[smi] device_rid: ResourceId,
+  #[string] label: Cow<str>,
+  #[serde] layout: GPUPipelineLayoutOrGPUAutoLayoutMode,
+  #[serde] compute: GpuProgrammableStage,
+) -> Result<WebGpuResult, AnyError> {
+  let instance = state.borrow::<super::Instance>();
+  let device_resource = state
+    .resource_table
+    .get::<super::WebGpuDevice>(device_rid)?;
+  let device = device_resource.1;
+
+  let pipeline_layout = match layout {
+    GPUPipelineLayoutOrGPUAutoLayoutMode::Layout(rid) => {
+      let id = state.resource_table.get::<WebGpuPipelineLayout>(rid)?;
+      Some(id.1)
+    }
+    GPUPipelineLayoutOrGPUAutoLayoutMode::Auto(GPUAutoLayoutMode::Auto) => None,
+  };
+
+  let compute_shader_module_resource =
+    state
+      .resource_table
+      .get::<super::shader::WebGpuShaderModule>(compute.module)?;
+
+  let descriptor = wgpu_core::pipeline::ComputePipelineDescriptor {
+    label: Some(label),
+    layout: pipeline_layout,
+    stage: wgpu_core::pipeline::ProgrammableStageDescriptor {
+      module: compute_shader_module_resource.1,
+      entry_point: Cow::from(compute.entry_point),
+      // TODO(lucacasonato): support args.compute.constants
+    },
+  };
+  let implicit_pipelines = match layout {
+    GPUPipelineLayoutOrGPUAutoLayoutMode::Layout(_) => None,
+    GPUPipelineLayoutOrGPUAutoLayoutMode::Auto(GPUAutoLayoutMode::Auto) => {
+      Some(wgpu_core::device::ImplicitPipelineIds {
+        root_id: (),
+        group_ids: &[(); MAX_BIND_GROUPS],
+      })
+    }
+  };
+
+  let (compute_pipeline, maybe_err) = gfx_select!(device => instance.device_create_compute_pipeline(
+    device,
+    &descriptor,
+    (),
+    implicit_pipelines
+  ));
+
+  let rid = state
+    .resource_table
+    .add(WebGpuComputePipeline(instance.clone(), compute_pipeline));
+
+  Ok(WebGpuResult::rid_err(rid, maybe_err))
+}
+
+#[derive(Serialize)]
+#[serde(rename_all = "camelCase")]
+pub struct PipelineLayout {
+  rid: ResourceId,
+  label: String,
+  err: Option<WebGpuError>,
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_compute_pipeline_get_bind_group_layout(
+  state: &mut OpState,
+  #[smi] compute_pipeline_rid: ResourceId,
+  index: u32,
+) -> Result<PipelineLayout, AnyError> {
+  let instance = state.borrow::<super::Instance>();
+  let compute_pipeline_resource = state
+    .resource_table
+    .get::<WebGpuComputePipeline>(compute_pipeline_rid)?;
+  let compute_pipeline = compute_pipeline_resource.1;
+
+  let (bind_group_layout, maybe_err) = gfx_select!(compute_pipeline => instance.compute_pipeline_get_bind_group_layout(compute_pipeline, index, ()));
+
+  let label = gfx_select!(bind_group_layout => instance.bind_group_layout_label(bind_group_layout));
+
+  let rid = state
+    .resource_table
+    .add(super::binding::WebGpuBindGroupLayout(
+      instance.clone(),
+      bind_group_layout,
+    ));
+
+  Ok(PipelineLayout {
+    rid,
+    label,
+    err: maybe_err.map(WebGpuError::from),
+  })
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "kebab-case")]
+pub enum GpuCullMode {
+  None,
+  Front,
+  Back,
+}
+
+impl From<GpuCullMode> for Option<wgpu_types::Face> {
+  fn from(value: GpuCullMode) -> Option<wgpu_types::Face> {
+    match value {
+      GpuCullMode::None => None,
+      GpuCullMode::Front => Some(wgpu_types::Face::Front),
+      GpuCullMode::Back => Some(wgpu_types::Face::Back),
+    }
+  }
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "camelCase")]
+struct GpuPrimitiveState {
+  topology: wgpu_types::PrimitiveTopology,
+  strip_index_format: Option<wgpu_types::IndexFormat>,
+  front_face: wgpu_types::FrontFace,
+  cull_mode: GpuCullMode,
+  unclipped_depth: bool,
+}
+
+impl From<GpuPrimitiveState> for wgpu_types::PrimitiveState {
+  fn from(value: GpuPrimitiveState) -> wgpu_types::PrimitiveState {
+    wgpu_types::PrimitiveState {
+      topology: value.topology,
+      strip_index_format: value.strip_index_format,
+      front_face: value.front_face,
+      cull_mode: value.cull_mode.into(),
+      unclipped_depth: value.unclipped_depth,
+      polygon_mode: Default::default(), // native-only
+      conservative: false,              // native-only
+    }
+  }
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "camelCase")]
+struct GpuDepthStencilState {
+  format: wgpu_types::TextureFormat,
+  depth_write_enabled: bool,
+  depth_compare: wgpu_types::CompareFunction,
+  stencil_front: wgpu_types::StencilFaceState,
+  stencil_back: wgpu_types::StencilFaceState,
+  stencil_read_mask: u32,
+  stencil_write_mask: u32,
+  depth_bias: i32,
+  depth_bias_slope_scale: f32,
+  depth_bias_clamp: f32,
+}
+
+impl From<GpuDepthStencilState> for wgpu_types::DepthStencilState {
+  fn from(state: GpuDepthStencilState) -> wgpu_types::DepthStencilState {
+    wgpu_types::DepthStencilState {
+      format: state.format,
+      depth_write_enabled: state.depth_write_enabled,
+      depth_compare: state.depth_compare,
+      stencil: wgpu_types::StencilState {
+        front: state.stencil_front,
+        back: state.stencil_back,
+        read_mask: state.stencil_read_mask,
+        write_mask: state.stencil_write_mask,
+      },
+      bias: wgpu_types::DepthBiasState {
+        constant: state.depth_bias,
+        slope_scale: state.depth_bias_slope_scale,
+        clamp: state.depth_bias_clamp,
+      },
+    }
+  }
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "camelCase")]
+struct GpuVertexBufferLayout {
+  array_stride: u64,
+  step_mode: wgpu_types::VertexStepMode,
+  attributes: Vec<wgpu_types::VertexAttribute>,
+}
+
+impl<'a> From<GpuVertexBufferLayout>
+  for wgpu_core::pipeline::VertexBufferLayout<'a>
+{
+  fn from(
+    layout: GpuVertexBufferLayout,
+  ) -> wgpu_core::pipeline::VertexBufferLayout<'a> {
+    wgpu_core::pipeline::VertexBufferLayout {
+      array_stride: layout.array_stride,
+      step_mode: layout.step_mode,
+      attributes: Cow::Owned(layout.attributes),
+    }
+  }
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "camelCase")]
+struct GpuVertexState {
+  module: ResourceId,
+  entry_point: String,
+  buffers: Vec<Option<GpuVertexBufferLayout>>,
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "camelCase")]
+struct GpuMultisampleState {
+  count: u32,
+  mask: u64,
+  alpha_to_coverage_enabled: bool,
+}
+
+impl From<GpuMultisampleState> for wgpu_types::MultisampleState {
+  fn from(gms: GpuMultisampleState) -> wgpu_types::MultisampleState {
+    wgpu_types::MultisampleState {
+      count: gms.count,
+      mask: gms.mask,
+      alpha_to_coverage_enabled: gms.alpha_to_coverage_enabled,
+    }
+  }
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "camelCase")]
+struct GpuFragmentState {
+  targets: Vec<Option<wgpu_types::ColorTargetState>>,
+  module: u32,
+  entry_point: String,
+  // TODO(lucacasonato): constants
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "camelCase")]
+pub struct CreateRenderPipelineArgs {
+  device_rid: ResourceId,
+  label: String,
+  layout: GPUPipelineLayoutOrGPUAutoLayoutMode,
+  vertex: GpuVertexState,
+  primitive: GpuPrimitiveState,
+  depth_stencil: Option<GpuDepthStencilState>,
+  multisample: wgpu_types::MultisampleState,
+  fragment: Option<GpuFragmentState>,
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_create_render_pipeline(
+  state: &mut OpState,
+  #[serde] args: CreateRenderPipelineArgs,
+) -> Result<WebGpuResult, AnyError> {
+  let instance = state.borrow::<super::Instance>();
+  let device_resource = state
+    .resource_table
+    .get::<super::WebGpuDevice>(args.device_rid)?;
+  let device = device_resource.1;
+
+  let layout = match args.layout {
+    GPUPipelineLayoutOrGPUAutoLayoutMode::Layout(rid) => {
+      let pipeline_layout_resource =
+        state.resource_table.get::<WebGpuPipelineLayout>(rid)?;
+      Some(pipeline_layout_resource.1)
+    }
+    GPUPipelineLayoutOrGPUAutoLayoutMode::Auto(GPUAutoLayoutMode::Auto) => None,
+  };
+
+  let vertex_shader_module_resource =
+    state
+      .resource_table
+      .get::<super::shader::WebGpuShaderModule>(args.vertex.module)?;
+
+  let fragment = if let Some(fragment) = args.fragment {
+    let fragment_shader_module_resource =
+      state
+        .resource_table
+        .get::<super::shader::WebGpuShaderModule>(fragment.module)?;
+
+    Some(wgpu_core::pipeline::FragmentState {
+      stage: wgpu_core::pipeline::ProgrammableStageDescriptor {
+        module: fragment_shader_module_resource.1,
+        entry_point: Cow::from(fragment.entry_point),
+      },
+      targets: Cow::from(fragment.targets),
+    })
+  } else {
+    None
+  };
+
+  let vertex_buffers = args
+    .vertex
+    .buffers
+    .into_iter()
+    .flatten()
+    .map(Into::into)
+    .collect();
+
+  let descriptor = wgpu_core::pipeline::RenderPipelineDescriptor {
+    label: Some(Cow::Owned(args.label)),
+    layout,
+    vertex: wgpu_core::pipeline::VertexState {
+      stage: wgpu_core::pipeline::ProgrammableStageDescriptor {
+        module: vertex_shader_module_resource.1,
+        entry_point: Cow::Owned(args.vertex.entry_point),
+      },
+      buffers: Cow::Owned(vertex_buffers),
+    },
+    primitive: args.primitive.into(),
+    depth_stencil: args.depth_stencil.map(Into::into),
+    multisample: args.multisample,
+    fragment,
+    multiview: None,
+  };
+
+  let implicit_pipelines = match args.layout {
+    GPUPipelineLayoutOrGPUAutoLayoutMode::Layout(_) => None,
+    GPUPipelineLayoutOrGPUAutoLayoutMode::Auto(GPUAutoLayoutMode::Auto) => {
+      Some(wgpu_core::device::ImplicitPipelineIds {
+        root_id: (),
+        group_ids: &[(); MAX_BIND_GROUPS],
+      })
+    }
+  };
+
+  let (render_pipeline, maybe_err) = gfx_select!(device => instance.device_create_render_pipeline(
+    device,
+    &descriptor,
+    (),
+    implicit_pipelines
+  ));
+
+  let rid = state
+    .resource_table
+    .add(WebGpuRenderPipeline(instance.clone(), render_pipeline));
+
+  Ok(WebGpuResult::rid_err(rid, maybe_err))
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_render_pipeline_get_bind_group_layout(
+  state: &mut OpState,
+  #[smi] render_pipeline_rid: ResourceId,
+  index: u32,
+) -> Result<PipelineLayout, AnyError> {
+  let instance = state.borrow::<super::Instance>();
+  let render_pipeline_resource = state
+    .resource_table
+    .get::<WebGpuRenderPipeline>(render_pipeline_rid)?;
+  let render_pipeline = render_pipeline_resource.1;
+
+  let (bind_group_layout, maybe_err) = gfx_select!(render_pipeline => instance.render_pipeline_get_bind_group_layout(render_pipeline, index, ()));
+
+  let label = gfx_select!(bind_group_layout => instance.bind_group_layout_label(bind_group_layout));
+
+  let rid = state
+    .resource_table
+    .add(super::binding::WebGpuBindGroupLayout(
+      instance.clone(),
+      bind_group_layout,
+    ));
+
+  Ok(PipelineLayout {
+    rid,
+    label,
+    err: maybe_err.map(WebGpuError::from),
+  })
+}

ext/webgpu/queue.rs

@@ -0,0 +1,131 @@
+// Copyright 2018-2023 the Deno authors. All rights reserved. MIT license.
+
+use crate::command_encoder::WebGpuCommandBuffer;
+use deno_core::error::AnyError;
+use deno_core::op2;
+use deno_core::OpState;
+use deno_core::Resource;
+use deno_core::ResourceId;
+use serde::Deserialize;
+
+use super::error::WebGpuResult;
+
+type WebGpuQueue = super::WebGpuDevice;
+
+#[op2]
+#[serde]
+pub fn op_webgpu_queue_submit(
+  state: &mut OpState,
+  #[smi] queue_rid: ResourceId,
+  #[serde] command_buffers: Vec<ResourceId>,
+) -> Result<WebGpuResult, AnyError> {
+  let instance = state.borrow::<super::Instance>();
+  let queue_resource = state.resource_table.get::<WebGpuQueue>(queue_rid)?;
+  let queue = queue_resource.1;
+
+  let ids = command_buffers
+    .iter()
+    .map(|rid| {
+      let buffer_resource =
+        state.resource_table.get::<WebGpuCommandBuffer>(*rid)?;
+      let mut id = buffer_resource.1.borrow_mut();
+      Ok(id.take().unwrap())
+    })
+    .collect::<Result<Vec<_>, AnyError>>()?;
+
+  let maybe_err =
+    gfx_select!(queue => instance.queue_submit(queue, &ids)).err();
+
+  for rid in command_buffers {
+    let resource = state.resource_table.take::<WebGpuCommandBuffer>(rid)?;
+    resource.close();
+  }
+
+  Ok(WebGpuResult::maybe_err(maybe_err))
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "camelCase")]
+pub struct GpuImageDataLayout {
+  offset: u64,
+  bytes_per_row: Option<u32>,
+  rows_per_image: Option<u32>,
+}
+
+impl From<GpuImageDataLayout> for wgpu_types::ImageDataLayout {
+  fn from(layout: GpuImageDataLayout) -> Self {
+    wgpu_types::ImageDataLayout {
+      offset: layout.offset,
+      bytes_per_row: layout.bytes_per_row,
+      rows_per_image: layout.rows_per_image,
+    }
+  }
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_write_buffer(
+  state: &mut OpState,
+  #[smi] queue_rid: ResourceId,
+  #[smi] buffer: ResourceId,
+  #[number] buffer_offset: u64,
+  #[number] data_offset: usize,
+  #[number] size: Option<usize>,
+  #[buffer] buf: &[u8],
+) -> Result<WebGpuResult, AnyError> {
+  let instance = state.borrow::<super::Instance>();
+  let buffer_resource = state
+    .resource_table
+    .get::<super::buffer::WebGpuBuffer>(buffer)?;
+  let buffer = buffer_resource.1;
+  let queue_resource = state.resource_table.get::<WebGpuQueue>(queue_rid)?;
+  let queue = queue_resource.1;
+
+  let data = match size {
+    Some(size) => &buf[data_offset..(data_offset + size)],
+    None => &buf[data_offset..],
+  };
+  let maybe_err = gfx_select!(queue => instance.queue_write_buffer(
+    queue,
+    buffer,
+    buffer_offset,
+    data
+  ))
+  .err();
+
+  Ok(WebGpuResult::maybe_err(maybe_err))
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_write_texture(
+  state: &mut OpState,
+  #[smi] queue_rid: ResourceId,
+  #[serde] destination: super::command_encoder::GpuImageCopyTexture,
+  #[serde] data_layout: GpuImageDataLayout,
+  #[serde] size: wgpu_types::Extent3d,
+  #[buffer] buf: &[u8],
+) -> Result<WebGpuResult, AnyError> {
+  let instance = state.borrow::<super::Instance>();
+  let texture_resource = state
+    .resource_table
+    .get::<super::texture::WebGpuTexture>(destination.texture)?;
+  let queue_resource = state.resource_table.get::<WebGpuQueue>(queue_rid)?;
+  let queue = queue_resource.1;
+
+  let destination = wgpu_core::command::ImageCopyTexture {
+    texture: texture_resource.id,
+    mip_level: destination.mip_level,
+    origin: destination.origin,
+    aspect: destination.aspect,
+  };
+  let data_layout = data_layout.into();
+
+  gfx_ok!(queue => instance.queue_write_texture(
+    queue,
+    &destination,
+    buf,
+    &data_layout,
+    &size
+  ))
+}

ext/webgpu/render_pass.rs

@@ -0,0 +1,519 @@
+// Copyright 2018-2023 the Deno authors. All rights reserved. MIT license.
+
+use deno_core::error::type_error;
+use deno_core::error::AnyError;
+use deno_core::op2;
+use deno_core::OpState;
+use deno_core::Resource;
+use deno_core::ResourceId;
+use serde::Deserialize;
+use std::borrow::Cow;
+use std::cell::RefCell;
+
+use super::error::WebGpuResult;
+
+pub(crate) struct WebGpuRenderPass(
+  pub(crate) RefCell<wgpu_core::command::RenderPass>,
+);
+impl Resource for WebGpuRenderPass {
+  fn name(&self) -> Cow<str> {
+    "webGPURenderPass".into()
+  }
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "camelCase")]
+pub struct RenderPassSetViewportArgs {
+  render_pass_rid: ResourceId,
+  x: f32,
+  y: f32,
+  width: f32,
+  height: f32,
+  min_depth: f32,
+  max_depth: f32,
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_render_pass_set_viewport(
+  state: &mut OpState,
+  #[serde] args: RenderPassSetViewportArgs,
+) -> Result<WebGpuResult, AnyError> {
+  let render_pass_resource = state
+    .resource_table
+    .get::<WebGpuRenderPass>(args.render_pass_rid)?;
+
+  wgpu_core::command::render_ffi::wgpu_render_pass_set_viewport(
+    &mut render_pass_resource.0.borrow_mut(),
+    args.x,
+    args.y,
+    args.width,
+    args.height,
+    args.min_depth,
+    args.max_depth,
+  );
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_render_pass_set_scissor_rect(
+  state: &mut OpState,
+  #[smi] render_pass_rid: ResourceId,
+  x: u32,
+  y: u32,
+  width: u32,
+  height: u32,
+) -> Result<WebGpuResult, AnyError> {
+  let render_pass_resource = state
+    .resource_table
+    .get::<WebGpuRenderPass>(render_pass_rid)?;
+
+  wgpu_core::command::render_ffi::wgpu_render_pass_set_scissor_rect(
+    &mut render_pass_resource.0.borrow_mut(),
+    x,
+    y,
+    width,
+    height,
+  );
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_render_pass_set_blend_constant(
+  state: &mut OpState,
+  #[smi] render_pass_rid: ResourceId,
+  #[serde] color: wgpu_types::Color,
+) -> Result<WebGpuResult, AnyError> {
+  let render_pass_resource = state
+    .resource_table
+    .get::<WebGpuRenderPass>(render_pass_rid)?;
+
+  wgpu_core::command::render_ffi::wgpu_render_pass_set_blend_constant(
+    &mut render_pass_resource.0.borrow_mut(),
+    &color,
+  );
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_render_pass_set_stencil_reference(
+  state: &mut OpState,
+  #[smi] render_pass_rid: ResourceId,
+  reference: u32,
+) -> Result<WebGpuResult, AnyError> {
+  let render_pass_resource = state
+    .resource_table
+    .get::<WebGpuRenderPass>(render_pass_rid)?;
+
+  wgpu_core::command::render_ffi::wgpu_render_pass_set_stencil_reference(
+    &mut render_pass_resource.0.borrow_mut(),
+    reference,
+  );
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_render_pass_begin_occlusion_query(
+  state: &mut OpState,
+  #[smi] render_pass_rid: ResourceId,
+  query_index: u32,
+) -> Result<WebGpuResult, AnyError> {
+  let render_pass_resource = state
+    .resource_table
+    .get::<WebGpuRenderPass>(render_pass_rid)?;
+
+  wgpu_core::command::render_ffi::wgpu_render_pass_begin_occlusion_query(
+    &mut render_pass_resource.0.borrow_mut(),
+    query_index,
+  );
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_render_pass_end_occlusion_query(
+  state: &mut OpState,
+  #[smi] render_pass_rid: ResourceId,
+) -> Result<WebGpuResult, AnyError> {
+  let render_pass_resource = state
+    .resource_table
+    .get::<WebGpuRenderPass>(render_pass_rid)?;
+
+  wgpu_core::command::render_ffi::wgpu_render_pass_end_occlusion_query(
+    &mut render_pass_resource.0.borrow_mut(),
+  );
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_render_pass_execute_bundles(
+  state: &mut OpState,
+  #[smi] render_pass_rid: ResourceId,
+  #[serde] bundles: Vec<u32>,
+) -> Result<WebGpuResult, AnyError> {
+  let bundles = bundles
+    .iter()
+    .map(|rid| {
+      let render_bundle_resource =
+        state
+          .resource_table
+          .get::<super::bundle::WebGpuRenderBundle>(*rid)?;
+      Ok(render_bundle_resource.1)
+    })
+    .collect::<Result<Vec<_>, AnyError>>()?;
+
+  let render_pass_resource = state
+    .resource_table
+    .get::<WebGpuRenderPass>(render_pass_rid)?;
+
+  // SAFETY: the raw pointer and length are of the same slice, and that slice
+  // lives longer than the below function invocation.
+  unsafe {
+    wgpu_core::command::render_ffi::wgpu_render_pass_execute_bundles(
+      &mut render_pass_resource.0.borrow_mut(),
+      bundles.as_ptr(),
+      bundles.len(),
+    );
+  }
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_render_pass_end(
+  state: &mut OpState,
+  #[smi] command_encoder_rid: ResourceId,
+  #[smi] render_pass_rid: ResourceId,
+) -> Result<WebGpuResult, AnyError> {
+  let command_encoder_resource = state
+    .resource_table
+    .get::<super::command_encoder::WebGpuCommandEncoder>(
+    command_encoder_rid,
+  )?;
+  let command_encoder = command_encoder_resource.1;
+  let render_pass_resource = state
+    .resource_table
+    .take::<WebGpuRenderPass>(render_pass_rid)?;
+  let render_pass = &render_pass_resource.0.borrow();
+  let instance = state.borrow::<super::Instance>();
+
+  gfx_ok!(command_encoder => instance.command_encoder_run_render_pass(command_encoder, render_pass))
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_render_pass_set_bind_group(
+  state: &mut OpState,
+  #[smi] render_pass_rid: ResourceId,
+  index: u32,
+  bind_group: u32,
+  #[buffer] dynamic_offsets_data: &[u32],
+  #[number] dynamic_offsets_data_start: usize,
+  #[number] dynamic_offsets_data_length: usize,
+) -> Result<WebGpuResult, AnyError> {
+  let bind_group_resource =
+    state
+      .resource_table
+      .get::<super::binding::WebGpuBindGroup>(bind_group)?;
+  let render_pass_resource = state
+    .resource_table
+    .get::<WebGpuRenderPass>(render_pass_rid)?;
+
+  let start = dynamic_offsets_data_start;
+  let len = dynamic_offsets_data_length;
+
+  // Assert that length and start are both in bounds
+  assert!(start <= dynamic_offsets_data.len());
+  assert!(len <= dynamic_offsets_data.len() - start);
+
+  let dynamic_offsets_data: &[u32] = &dynamic_offsets_data[start..start + len];
+
+  // SAFETY: the raw pointer and length are of the same slice, and that slice
+  // lives longer than the below function invocation.
+  unsafe {
+    wgpu_core::command::render_ffi::wgpu_render_pass_set_bind_group(
+      &mut render_pass_resource.0.borrow_mut(),
+      index,
+      bind_group_resource.1,
+      dynamic_offsets_data.as_ptr(),
+      dynamic_offsets_data.len(),
+    );
+  }
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_render_pass_push_debug_group(
+  state: &mut OpState,
+  #[smi] render_pass_rid: ResourceId,
+  #[string] group_label: &str,
+) -> Result<WebGpuResult, AnyError> {
+  let render_pass_resource = state
+    .resource_table
+    .get::<WebGpuRenderPass>(render_pass_rid)?;
+
+  let label = std::ffi::CString::new(group_label).unwrap();
+  // SAFETY: the string the raw pointer points to lives longer than the below
+  // function invocation.
+  unsafe {
+    wgpu_core::command::render_ffi::wgpu_render_pass_push_debug_group(
+      &mut render_pass_resource.0.borrow_mut(),
+      label.as_ptr(),
+      0, // wgpu#975
+    );
+  }
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_render_pass_pop_debug_group(
+  state: &mut OpState,
+  #[smi] render_pass_rid: ResourceId,
+) -> Result<WebGpuResult, AnyError> {
+  let render_pass_resource = state
+    .resource_table
+    .get::<WebGpuRenderPass>(render_pass_rid)?;
+
+  wgpu_core::command::render_ffi::wgpu_render_pass_pop_debug_group(
+    &mut render_pass_resource.0.borrow_mut(),
+  );
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_render_pass_insert_debug_marker(
+  state: &mut OpState,
+  #[smi] render_pass_rid: ResourceId,
+  #[string] marker_label: &str,
+) -> Result<WebGpuResult, AnyError> {
+  let render_pass_resource = state
+    .resource_table
+    .get::<WebGpuRenderPass>(render_pass_rid)?;
+
+  let label = std::ffi::CString::new(marker_label).unwrap();
+  // SAFETY: the string the raw pointer points to lives longer than the below
+  // function invocation.
+  unsafe {
+    wgpu_core::command::render_ffi::wgpu_render_pass_insert_debug_marker(
+      &mut render_pass_resource.0.borrow_mut(),
+      label.as_ptr(),
+      0, // wgpu#975
+    );
+  }
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_render_pass_set_pipeline(
+  state: &mut OpState,
+  #[smi] render_pass_rid: ResourceId,
+  pipeline: u32,
+) -> Result<WebGpuResult, AnyError> {
+  let render_pipeline_resource =
+    state
+      .resource_table
+      .get::<super::pipeline::WebGpuRenderPipeline>(pipeline)?;
+  let render_pass_resource = state
+    .resource_table
+    .get::<WebGpuRenderPass>(render_pass_rid)?;
+
+  wgpu_core::command::render_ffi::wgpu_render_pass_set_pipeline(
+    &mut render_pass_resource.0.borrow_mut(),
+    render_pipeline_resource.1,
+  );
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_render_pass_set_index_buffer(
+  state: &mut OpState,
+  #[smi] render_pass_rid: ResourceId,
+  buffer: u32,
+  #[serde] index_format: wgpu_types::IndexFormat,
+  #[number] offset: u64,
+  #[number] size: Option<u64>,
+) -> Result<WebGpuResult, AnyError> {
+  let buffer_resource = state
+    .resource_table
+    .get::<super::buffer::WebGpuBuffer>(buffer)?;
+  let render_pass_resource = state
+    .resource_table
+    .get::<WebGpuRenderPass>(render_pass_rid)?;
+
+  let size = if let Some(size) = size {
+    Some(
+      std::num::NonZeroU64::new(size)
+        .ok_or_else(|| type_error("size must be larger than 0"))?,
+    )
+  } else {
+    None
+  };
+
+  render_pass_resource.0.borrow_mut().set_index_buffer(
+    buffer_resource.1,
+    index_format,
+    offset,
+    size,
+  );
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_render_pass_set_vertex_buffer(
+  state: &mut OpState,
+  #[smi] render_pass_rid: ResourceId,
+  slot: u32,
+  buffer: u32,
+  #[number] offset: u64,
+  #[number] size: Option<u64>,
+) -> Result<WebGpuResult, AnyError> {
+  let buffer_resource = state
+    .resource_table
+    .get::<super::buffer::WebGpuBuffer>(buffer)?;
+  let render_pass_resource = state
+    .resource_table
+    .get::<WebGpuRenderPass>(render_pass_rid)?;
+
+  let size = if let Some(size) = size {
+    Some(
+      std::num::NonZeroU64::new(size)
+        .ok_or_else(|| type_error("size must be larger than 0"))?,
+    )
+  } else {
+    None
+  };
+
+  wgpu_core::command::render_ffi::wgpu_render_pass_set_vertex_buffer(
+    &mut render_pass_resource.0.borrow_mut(),
+    slot,
+    buffer_resource.1,
+    offset,
+    size,
+  );
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_render_pass_draw(
+  state: &mut OpState,
+  #[smi] render_pass_rid: ResourceId,
+  vertex_count: u32,
+  instance_count: u32,
+  first_vertex: u32,
+  first_instance: u32,
+) -> Result<WebGpuResult, AnyError> {
+  let render_pass_resource = state
+    .resource_table
+    .get::<WebGpuRenderPass>(render_pass_rid)?;
+
+  wgpu_core::command::render_ffi::wgpu_render_pass_draw(
+    &mut render_pass_resource.0.borrow_mut(),
+    vertex_count,
+    instance_count,
+    first_vertex,
+    first_instance,
+  );
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_render_pass_draw_indexed(
+  state: &mut OpState,
+  #[smi] render_pass_rid: ResourceId,
+  index_count: u32,
+  instance_count: u32,
+  first_index: u32,
+  base_vertex: i32,
+  first_instance: u32,
+) -> Result<WebGpuResult, AnyError> {
+  let render_pass_resource = state
+    .resource_table
+    .get::<WebGpuRenderPass>(render_pass_rid)?;
+
+  wgpu_core::command::render_ffi::wgpu_render_pass_draw_indexed(
+    &mut render_pass_resource.0.borrow_mut(),
+    index_count,
+    instance_count,
+    first_index,
+    base_vertex,
+    first_instance,
+  );
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_render_pass_draw_indirect(
+  state: &mut OpState,
+  #[smi] render_pass_rid: ResourceId,
+  indirect_buffer: u32,
+  #[number] indirect_offset: u64,
+) -> Result<WebGpuResult, AnyError> {
+  let buffer_resource = state
+    .resource_table
+    .get::<super::buffer::WebGpuBuffer>(indirect_buffer)?;
+  let render_pass_resource = state
+    .resource_table
+    .get::<WebGpuRenderPass>(render_pass_rid)?;
+
+  wgpu_core::command::render_ffi::wgpu_render_pass_draw_indirect(
+    &mut render_pass_resource.0.borrow_mut(),
+    buffer_resource.1,
+    indirect_offset,
+  );
+
+  Ok(WebGpuResult::empty())
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_render_pass_draw_indexed_indirect(
+  state: &mut OpState,
+  #[smi] render_pass_rid: ResourceId,
+  indirect_buffer: u32,
+  #[number] indirect_offset: u64,
+) -> Result<WebGpuResult, AnyError> {
+  let buffer_resource = state
+    .resource_table
+    .get::<super::buffer::WebGpuBuffer>(indirect_buffer)?;
+  let render_pass_resource = state
+    .resource_table
+    .get::<WebGpuRenderPass>(render_pass_rid)?;
+
+  wgpu_core::command::render_ffi::wgpu_render_pass_draw_indexed_indirect(
+    &mut render_pass_resource.0.borrow_mut(),
+    buffer_resource.1,
+    indirect_offset,
+  );
+
+  Ok(WebGpuResult::empty())
+}

ext/webgpu/sampler.rs

@@ -0,0 +1,80 @@
+// Copyright 2018-2023 the Deno authors. All rights reserved. MIT license.
+
+use deno_core::error::AnyError;
+use deno_core::op2;
+use deno_core::OpState;
+use deno_core::Resource;
+use deno_core::ResourceId;
+use serde::Deserialize;
+use std::borrow::Cow;
+use std::rc::Rc;
+
+use super::error::WebGpuResult;
+
+pub(crate) struct WebGpuSampler(
+  pub(crate) crate::Instance,
+  pub(crate) wgpu_core::id::SamplerId,
+);
+impl Resource for WebGpuSampler {
+  fn name(&self) -> Cow<str> {
+    "webGPUSampler".into()
+  }
+
+  fn close(self: Rc<Self>) {
+    let instance = &self.0;
+    gfx_select!(self.1 => instance.sampler_drop(self.1));
+  }
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "camelCase")]
+pub struct CreateSamplerArgs {
+  device_rid: ResourceId,
+  label: String,
+  address_mode_u: wgpu_types::AddressMode,
+  address_mode_v: wgpu_types::AddressMode,
+  address_mode_w: wgpu_types::AddressMode,
+  mag_filter: wgpu_types::FilterMode,
+  min_filter: wgpu_types::FilterMode,
+  mipmap_filter: wgpu_types::FilterMode, // TODO: GPUMipmapFilterMode
+  lod_min_clamp: f32,
+  lod_max_clamp: f32,
+  compare: Option<wgpu_types::CompareFunction>,
+  max_anisotropy: u16,
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_create_sampler(
+  state: &mut OpState,
+  #[serde] args: CreateSamplerArgs,
+) -> Result<WebGpuResult, AnyError> {
+  let instance = state.borrow::<super::Instance>();
+  let device_resource = state
+    .resource_table
+    .get::<super::WebGpuDevice>(args.device_rid)?;
+  let device = device_resource.1;
+
+  let descriptor = wgpu_core::resource::SamplerDescriptor {
+    label: Some(Cow::Owned(args.label)),
+    address_modes: [
+      args.address_mode_u,
+      args.address_mode_v,
+      args.address_mode_w,
+    ],
+    mag_filter: args.mag_filter,
+    min_filter: args.min_filter,
+    mipmap_filter: args.mipmap_filter,
+    lod_min_clamp: args.lod_min_clamp,
+    lod_max_clamp: args.lod_max_clamp,
+    compare: args.compare,
+    anisotropy_clamp: args.max_anisotropy,
+    border_color: None, // native-only
+  };
+
+  gfx_put!(device => instance.device_create_sampler(
+    device,
+    &descriptor,
+    ()
+  ) => state, WebGpuSampler)
+}

ext/webgpu/shader.rs

@@ -0,0 +1,55 @@
+// Copyright 2018-2023 the Deno authors. All rights reserved. MIT license.
+
+use deno_core::error::AnyError;
+use deno_core::op2;
+use deno_core::OpState;
+use deno_core::Resource;
+use deno_core::ResourceId;
+use std::borrow::Cow;
+use std::rc::Rc;
+
+use super::error::WebGpuResult;
+
+pub(crate) struct WebGpuShaderModule(
+  pub(crate) super::Instance,
+  pub(crate) wgpu_core::id::ShaderModuleId,
+);
+impl Resource for WebGpuShaderModule {
+  fn name(&self) -> Cow<str> {
+    "webGPUShaderModule".into()
+  }
+
+  fn close(self: Rc<Self>) {
+    let instance = &self.0;
+    gfx_select!(self.1 => instance.shader_module_drop(self.1));
+  }
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_create_shader_module(
+  state: &mut OpState,
+  #[smi] device_rid: ResourceId,
+  #[string] label: Cow<str>,
+  #[string] code: Cow<str>,
+) -> Result<WebGpuResult, AnyError> {
+  let instance = state.borrow::<super::Instance>();
+  let device_resource = state
+    .resource_table
+    .get::<super::WebGpuDevice>(device_rid)?;
+  let device = device_resource.1;
+
+  let source = wgpu_core::pipeline::ShaderModuleSource::Wgsl(code);
+
+  let descriptor = wgpu_core::pipeline::ShaderModuleDescriptor {
+    label: Some(label),
+    shader_bound_checks: wgpu_types::ShaderBoundChecks::default(),
+  };
+
+  gfx_put!(device => instance.device_create_shader_module(
+    device,
+    &descriptor,
+    source,
+    ()
+  ) => state, WebGpuShaderModule)
+}

ext/webgpu/surface.rs

@@ -0,0 +1,133 @@
+// Copyright 2018-2023 the Deno authors. All rights reserved. MIT license.
+
+use super::WebGpuResult;
+use deno_core::error::AnyError;
+use deno_core::op2;
+use deno_core::OpState;
+use deno_core::Resource;
+use deno_core::ResourceId;
+use serde::Deserialize;
+use std::borrow::Cow;
+use std::rc::Rc;
+use wgpu_types::SurfaceStatus;
+
+deno_core::extension!(
+  deno_webgpu_surface,
+  deps = [deno_webidl, deno_web, deno_webgpu],
+  ops = [
+    op_webgpu_surface_configure,
+    op_webgpu_surface_get_current_texture,
+    op_webgpu_surface_present,
+  ],
+  esm = ["02_surface.js"],
+);
+
+pub struct WebGpuSurface(pub crate::Instance, pub wgpu_core::id::SurfaceId);
+impl Resource for WebGpuSurface {
+  fn name(&self) -> Cow<str> {
+    "webGPUSurface".into()
+  }
+
+  fn close(self: Rc<Self>) {
+    self.0.surface_drop(self.1);
+  }
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "camelCase")]
+pub struct SurfaceConfigureArgs {
+  surface_rid: ResourceId,
+  device_rid: ResourceId,
+  format: wgpu_types::TextureFormat,
+  usage: u32,
+  width: u32,
+  height: u32,
+  present_mode: Option<wgpu_types::PresentMode>,
+  alpha_mode: wgpu_types::CompositeAlphaMode,
+  view_formats: Vec<wgpu_types::TextureFormat>,
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_surface_configure(
+  state: &mut OpState,
+  #[serde] args: SurfaceConfigureArgs,
+) -> Result<WebGpuResult, AnyError> {
+  let instance = state.borrow::<super::Instance>();
+  let device_resource = state
+    .resource_table
+    .get::<super::WebGpuDevice>(args.device_rid)?;
+  let device = device_resource.1;
+  let surface_resource = state
+    .resource_table
+    .get::<WebGpuSurface>(args.surface_rid)?;
+  let surface = surface_resource.1;
+
+  let conf = wgpu_types::SurfaceConfiguration::<Vec<wgpu_types::TextureFormat>> {
+    usage: wgpu_types::TextureUsages::from_bits_truncate(args.usage),
+    format: args.format,
+    width: args.width,
+    height: args.height,
+    present_mode: args.present_mode.unwrap_or_default(),
+    alpha_mode: args.alpha_mode,
+    view_formats: args.view_formats,
+  };
+
+  let err =
+    gfx_select!(device => instance.surface_configure(surface, device, &conf));
+
+  Ok(WebGpuResult::maybe_err(err))
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_surface_get_current_texture(
+  state: &mut OpState,
+  #[smi] device_rid: ResourceId,
+  #[smi] surface_rid: ResourceId,
+) -> Result<WebGpuResult, AnyError> {
+  let instance = state.borrow::<super::Instance>();
+  let device_resource = state
+    .resource_table
+    .get::<super::WebGpuDevice>(device_rid)?;
+  let device = device_resource.1;
+  let surface_resource =
+    state.resource_table.get::<WebGpuSurface>(surface_rid)?;
+  let surface = surface_resource.1;
+
+  let output =
+    gfx_select!(device => instance.surface_get_current_texture(surface, ()))?;
+
+  match output.status {
+    SurfaceStatus::Good | SurfaceStatus::Suboptimal => {
+      let id = output.texture_id.unwrap();
+      let rid = state.resource_table.add(crate::texture::WebGpuTexture {
+        instance: instance.clone(),
+        id,
+        owned: false,
+      });
+      Ok(WebGpuResult::rid(rid))
+    }
+    _ => Err(AnyError::msg("Invalid Surface Status")),
+  }
+}
+
+#[op2(fast)]
+pub fn op_webgpu_surface_present(
+  state: &mut OpState,
+  #[smi] device_rid: ResourceId,
+  #[smi] surface_rid: ResourceId,
+) -> Result<(), AnyError> {
+  let instance = state.borrow::<super::Instance>();
+  let device_resource = state
+    .resource_table
+    .get::<super::WebGpuDevice>(device_rid)?;
+  let device = device_resource.1;
+  let surface_resource =
+    state.resource_table.get::<WebGpuSurface>(surface_rid)?;
+  let surface = surface_resource.1;
+
+  let _ = gfx_select!(device => instance.surface_present(surface))?;
+
+  Ok(())
+}

ext/webgpu/texture.rs

@@ -0,0 +1,134 @@
+// Copyright 2018-2023 the Deno authors. All rights reserved. MIT license.
+
+use deno_core::error::AnyError;
+use deno_core::op2;
+use deno_core::OpState;
+use deno_core::Resource;
+use deno_core::ResourceId;
+use serde::Deserialize;
+use std::borrow::Cow;
+use std::rc::Rc;
+
+use super::error::WebGpuResult;
+pub(crate) struct WebGpuTexture {
+  pub(crate) instance: crate::Instance,
+  pub(crate) id: wgpu_core::id::TextureId,
+  pub(crate) owned: bool,
+}
+
+impl Resource for WebGpuTexture {
+  fn name(&self) -> Cow<str> {
+    "webGPUTexture".into()
+  }
+
+  fn close(self: Rc<Self>) {
+    if self.owned {
+      let instance = &self.instance;
+      gfx_select!(self.id => instance.texture_drop(self.id, true));
+    }
+  }
+}
+
+pub(crate) struct WebGpuTextureView(
+  pub(crate) crate::Instance,
+  pub(crate) wgpu_core::id::TextureViewId,
+);
+impl Resource for WebGpuTextureView {
+  fn name(&self) -> Cow<str> {
+    "webGPUTextureView".into()
+  }
+
+  fn close(self: Rc<Self>) {
+    let instance = &self.0;
+    gfx_select!(self.1 => instance.texture_view_drop(self.1, true)).unwrap();
+  }
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "camelCase")]
+pub struct CreateTextureArgs {
+  device_rid: ResourceId,
+  label: String,
+  size: wgpu_types::Extent3d,
+  mip_level_count: u32,
+  sample_count: u32,
+  dimension: wgpu_types::TextureDimension,
+  format: wgpu_types::TextureFormat,
+  usage: u32,
+  view_formats: Vec<wgpu_types::TextureFormat>,
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_create_texture(
+  state: &mut OpState,
+  #[serde] args: CreateTextureArgs,
+) -> Result<WebGpuResult, AnyError> {
+  let instance = state.borrow::<super::Instance>();
+  let device_resource = state
+    .resource_table
+    .get::<super::WebGpuDevice>(args.device_rid)?;
+  let device = device_resource.1;
+
+  let descriptor = wgpu_core::resource::TextureDescriptor {
+    label: Some(Cow::Owned(args.label)),
+    size: args.size,
+    mip_level_count: args.mip_level_count,
+    sample_count: args.sample_count,
+    dimension: args.dimension,
+    format: args.format,
+    usage: wgpu_types::TextureUsages::from_bits_truncate(args.usage),
+    view_formats: args.view_formats,
+  };
+
+  let (val, maybe_err) = gfx_select!(device => instance.device_create_texture(
+    device,
+    &descriptor,
+    ()
+  ));
+
+  let rid = state.resource_table.add(WebGpuTexture {
+    instance: instance.clone(),
+    id: val,
+    owned: true,
+  });
+
+  Ok(WebGpuResult::rid_err(rid, maybe_err))
+}
+
+#[derive(Deserialize)]
+#[serde(rename_all = "camelCase")]
+pub struct CreateTextureViewArgs {
+  texture_rid: ResourceId,
+  label: String,
+  format: Option<wgpu_types::TextureFormat>,
+  dimension: Option<wgpu_types::TextureViewDimension>,
+  #[serde(flatten)]
+  range: wgpu_types::ImageSubresourceRange,
+}
+
+#[op2]
+#[serde]
+pub fn op_webgpu_create_texture_view(
+  state: &mut OpState,
+  #[serde] args: CreateTextureViewArgs,
+) -> Result<WebGpuResult, AnyError> {
+  let instance = state.borrow::<super::Instance>();
+  let texture_resource = state
+    .resource_table
+    .get::<WebGpuTexture>(args.texture_rid)?;
+  let texture = texture_resource.id;
+
+  let descriptor = wgpu_core::resource::TextureViewDescriptor {
+    label: Some(Cow::Owned(args.label)),
+    format: args.format,
+    dimension: args.dimension,
+    range: args.range,
+  };
+
+  gfx_put!(texture => instance.texture_create_view(
+    texture,
+    &descriptor,
+    ()
+  ) => state, WebGpuTextureView)
+}

runtime/Cargo.toml

@@ -57,6 +57,7 @@ deno_kv.workspace = true
 deno_tls.workspace = true
 deno_url.workspace = true
 deno_web.workspace = true
+deno_webgpu.workspace = true
 deno_webidl.workspace = true
 deno_websocket.workspace = true
 deno_webstorage.workspace = true
@@ -88,6 +89,7 @@ deno_node.workspace = true
 deno_tls.workspace = true
 deno_url.workspace = true
 deno_web.workspace = true
+deno_webgpu.workspace = true
 deno_webidl.workspace = true
 deno_websocket.workspace = true
 deno_webstorage.workspace = true

runtime/build.rs

@@ -224,6 +224,7 @@ mod startup_snapshot {
         Default::default(),
         Default::default(),
       ),
+      deno_webgpu::deno_webgpu::init_ops_and_esm(),
       deno_fetch::deno_fetch::init_ops_and_esm::<Permissions>(
         Default::default(),
       ),

runtime/errors.rs

@@ -167,6 +167,7 @@ pub fn get_nix_error_class(error: &nix::Error) -> &'static str {
 
 pub fn get_error_class_name(e: &AnyError) -> Option<&'static str> {
   deno_core::error::get_custom_error_class(e)
+    .or_else(|| deno_webgpu::error::get_error_class_name(e))
     .or_else(|| deno_web::get_error_class_name(e))
     .or_else(|| deno_webstorage::get_not_supported_error_class_name(e))
     .or_else(|| deno_websocket::get_network_error_class_name(e))

runtime/js/90_deno_ns.js

@@ -164,7 +164,8 @@ const unstableIds = {
   kv: 6,
   net: 7,
   unsafeProto: 8,
-  workerOptions: 9,
+  webgpu: 9,
+  workerOptions: 10,
 };
 
 const denoNsUnstableById = {};
@@ -216,6 +217,8 @@ denoNsUnstableById[unstableIds.net] = {
 
 // denoNsUnstableById[unstableIds.unsafeProto] = {}
 
+// denoNsUnstableById[unstableIds.webgpu] = {}
+
 // denoNsUnstableById[unstableIds.workerOptions] = {}
 
 // when editing this list, also update unstableDenoProps in cli/tsc/99_main_compiler.js

runtime/js/98_global_scope.js

@@ -151,6 +151,45 @@ unstableForWindowOrWorkerGlobalScope[unstableIds.broadcastChannel] = {
 unstableForWindowOrWorkerGlobalScope[unstableIds.net] = {
   WebSocketStream: util.nonEnumerable(webSocketStream.WebSocketStream),
 };
+unstableForWindowOrWorkerGlobalScope[unstableIds.webgpu] = {
+  GPU: webGPUNonEnumerable(() => webgpu.GPU),
+  GPUAdapter: webGPUNonEnumerable(() => webgpu.GPUAdapter),
+  GPUAdapterInfo: webGPUNonEnumerable(() => webgpu.GPUAdapterInfo),
+  GPUSupportedLimits: webGPUNonEnumerable(() => webgpu.GPUSupportedLimits),
+  GPUSupportedFeatures: webGPUNonEnumerable(() => webgpu.GPUSupportedFeatures),
+  GPUDeviceLostInfo: webGPUNonEnumerable(() => webgpu.GPUDeviceLostInfo),
+  GPUDevice: webGPUNonEnumerable(() => webgpu.GPUDevice),
+  GPUQueue: webGPUNonEnumerable(() => webgpu.GPUQueue),
+  GPUBuffer: webGPUNonEnumerable(() => webgpu.GPUBuffer),
+  GPUBufferUsage: webGPUNonEnumerable(() => webgpu.GPUBufferUsage),
+  GPUMapMode: webGPUNonEnumerable(() => webgpu.GPUMapMode),
+  GPUTextureUsage: webGPUNonEnumerable(() => webgpu.GPUTextureUsage),
+  GPUTexture: webGPUNonEnumerable(() => webgpu.GPUTexture),
+  GPUTextureView: webGPUNonEnumerable(() => webgpu.GPUTextureView),
+  GPUSampler: webGPUNonEnumerable(() => webgpu.GPUSampler),
+  GPUBindGroupLayout: webGPUNonEnumerable(() => webgpu.GPUBindGroupLayout),
+  GPUPipelineLayout: webGPUNonEnumerable(() => webgpu.GPUPipelineLayout),
+  GPUBindGroup: webGPUNonEnumerable(() => webgpu.GPUBindGroup),
+  GPUShaderModule: webGPUNonEnumerable(() => webgpu.GPUShaderModule),
+  GPUShaderStage: webGPUNonEnumerable(() => webgpu.GPUShaderStage),
+  GPUComputePipeline: webGPUNonEnumerable(() => webgpu.GPUComputePipeline),
+  GPURenderPipeline: webGPUNonEnumerable(() => webgpu.GPURenderPipeline),
+  GPUColorWrite: webGPUNonEnumerable(() => webgpu.GPUColorWrite),
+  GPUCommandEncoder: webGPUNonEnumerable(() => webgpu.GPUCommandEncoder),
+  GPURenderPassEncoder: webGPUNonEnumerable(() => webgpu.GPURenderPassEncoder),
+  GPUComputePassEncoder: webGPUNonEnumerable(() =>
+    webgpu.GPUComputePassEncoder
+  ),
+  GPUCommandBuffer: webGPUNonEnumerable(() => webgpu.GPUCommandBuffer),
+  GPURenderBundleEncoder: webGPUNonEnumerable(() =>
+    webgpu.GPURenderBundleEncoder
+  ),
+  GPURenderBundle: webGPUNonEnumerable(() => webgpu.GPURenderBundle),
+  GPUQuerySet: webGPUNonEnumerable(() => webgpu.GPUQuerySet),
+  GPUError: webGPUNonEnumerable(() => webgpu.GPUError),
+  GPUValidationError: webGPUNonEnumerable(() => webgpu.GPUValidationError),
+  GPUOutOfMemoryError: webGPUNonEnumerable(() => webgpu.GPUOutOfMemoryError),
+};
 
 class Navigator {
   constructor() {
@@ -190,7 +229,49 @@ const numCpus = memoizeLazy(() => ops.op_bootstrap_numcpus());
 const userAgent = memoizeLazy(() => ops.op_bootstrap_user_agent());
 const language = memoizeLazy(() => ops.op_bootstrap_language());
 
+let webgpu;
+
+function webGPUNonEnumerable(getter) {
+  let valueIsSet = false;
+  let value;
+
+  return {
+    get() {
+      loadWebGPU();
+
+      if (valueIsSet) {
+        return value;
+      } else {
+        return getter();
+      }
+    },
+    set(v) {
+      loadWebGPU();
+
+      valueIsSet = true;
+      value = v;
+    },
+    enumerable: false,
+    configurable: true,
+  };
+}
+
+function loadWebGPU() {
+  if (!webgpu) {
+    webgpu = ops.op_lazy_load_esm("ext:deno_webgpu/01_webgpu.js");
+  }
+}
+
 ObjectDefineProperties(Navigator.prototype, {
+  gpu: {
+    configurable: true,
+    enumerable: true,
+    get() {
+      webidl.assertBranded(this, NavigatorPrototype);
+      loadWebGPU();
+      return webgpu.gpu;
+    },
+  },
   hardwareConcurrency: {
     configurable: true,
     enumerable: true,
@@ -251,6 +332,15 @@ class WorkerNavigator {
 const workerNavigator = webidl.createBranded(WorkerNavigator);
 
 ObjectDefineProperties(WorkerNavigator.prototype, {
+  gpu: {
+    configurable: true,
+    enumerable: true,
+    get() {
+      webidl.assertBranded(this, WorkerNavigatorPrototype);
+      loadWebGPU();
+      return webgpu.gpu;
+    },
+  },
   hardwareConcurrency: {
     configurable: true,
     enumerable: true,

runtime/lib.rs

@@ -18,6 +18,7 @@ pub use deno_node;
 pub use deno_tls;
 pub use deno_url;
 pub use deno_web;
+pub use deno_webgpu;
 pub use deno_webidl;
 pub use deno_websocket;
 pub use deno_webstorage;

runtime/snapshot.rs

@@ -203,6 +203,7 @@ pub fn create_runtime_snapshot(
       Default::default(),
       Default::default(),
     ),
+    deno_webgpu::deno_webgpu::init_ops_and_esm(),
     deno_fetch::deno_fetch::init_ops_and_esm::<Permissions>(Default::default()),
     deno_cache::deno_cache::init_ops_and_esm::<SqliteBackedCache>(None),
     deno_websocket::deno_websocket::init_ops_and_esm::<Permissions>(

runtime/web_worker.rs

@@ -408,6 +408,7 @@ impl WebWorker {
         options.blob_store.clone(),
         Some(main_module.clone()),
       ),
+      deno_webgpu::deno_webgpu::init_ops_and_esm(),
       deno_fetch::deno_fetch::init_ops_and_esm::<PermissionsContainer>(
         deno_fetch::Options {
           user_agent: options.bootstrap.user_agent.clone(),

runtime/worker.rs

@@ -307,6 +307,7 @@ impl MainWorker {
         options.blob_store.clone(),
         options.bootstrap.location.clone(),
       ),
+      deno_webgpu::deno_webgpu::init_ops_and_esm(),
       deno_fetch::deno_fetch::init_ops_and_esm::<PermissionsContainer>(
         deno_fetch::Options {
           user_agent: options.bootstrap.user_agent.clone(),

tools/README.md

@@ -31,6 +31,19 @@ executable
 cargo run -- run --allow-read --allow-write --allow-run --unstable ./tools/<script>
 ```
 
+## wgpu_sync.js
+
+`wgpu_sync.js` streamlines updating `deno_webgpu` from
+[gfx-rs/wgpu](https://github.com/gfx-rs/wgpu/).
+
+It essentially vendors the `deno_webgpu` tree with a few minor patches applied
+on top, somewhat similar to `git subtree`.
+
+1. Update `COMMIT` or `V_WGPU` in `./tools/wgpu_sync.js`
+2. Run `./tools/wgpu_sync.js`
+3. Double check changes, possibly patch
+4. Commit & send a PR with the updates
+
 ## copyright_checker.js
 
 `copyright_checker.js` is used to check copyright headers in the codebase.

tools/wgpu_sync.js

@@ -0,0 +1,92 @@
+#!/usr/bin/env -S deno run --unstable --allow-read --allow-write --allow-run
+// Copyright 2018-2023 the Deno authors. All rights reserved. MIT license.
+
+import { join, ROOT_PATH } from "./util.js";
+
+const COMMIT = "49b7ec97c164bac9ee877f45cdd806fbefecc5a4";
+const REPO = "gfx-rs/wgpu";
+const V_WGPU = "0.18";
+const TARGET_DIR = join(ROOT_PATH, "ext", "webgpu");
+
+async function bash(subcmd, opts = {}) {
+  const { success, code } = await new Deno.Command("bash", {
+    ...opts,
+    args: ["-c", subcmd],
+    stdout: "inherit",
+    sdterr: "inherit",
+  }).output();
+
+  // Exit process on failure
+  if (!success) {
+    Deno.exit(code);
+  }
+}
+
+async function clearTargetDir() {
+  await bash(`rm -r ${TARGET_DIR}/*`);
+}
+
+async function checkoutUpstream() {
+  // Path of deno_webgpu inside the TAR
+  const tarPrefix = `${REPO.replace("/", "-")}-${
+    COMMIT.slice(0, 7)
+  }/deno_webgpu/`;
+  const cmd =
+    `curl -L https://api.github.com/repos/${REPO}/tarball/${COMMIT} | tar -C '${TARGET_DIR}' -xzvf - --strip=2 '${tarPrefix}'`;
+  // console.log(cmd);
+  await bash(cmd);
+}
+
+async function denoWebgpuVersion() {
+  const coreCargo = join(ROOT_PATH, "Cargo.toml");
+  const contents = await Deno.readTextFile(coreCargo);
+  return contents.match(
+    /^deno_webgpu = { version = "(\d+\.\d+\.\d+)", path = ".\/ext\/webgpu" }$/m,
+  )[1];
+}
+
+async function patchFile(path, patcher) {
+  const data = await Deno.readTextFile(path);
+  const patched = patcher(data);
+  await Deno.writeTextFile(path, patched);
+}
+
+async function patchCargo() {
+  const vDenoWebgpu = await denoWebgpuVersion();
+  await patchFile(
+    join(TARGET_DIR, "Cargo.toml"),
+    (data) =>
+      data
+        .replace(/^version = .*/m, `version = "${vDenoWebgpu}"`)
+        .replace(
+          /^repository.workspace = true/m,
+          `repository = "https://github.com/gfx-rs/wgpu"`,
+        )
+        .replace(
+          /^serde = { workspace = true, features = ["derive"] }/m,
+          `serde.workspace = true`,
+        )
+        .replace(
+          /^tokio = { workspace = true, features = ["full"] }/m,
+          `tokio.workspace = true`,
+        ),
+  );
+
+  await patchFile(
+    join(ROOT_PATH, "Cargo.toml"),
+    (data) =>
+      data
+        .replace(/^wgpu-core = .*/m, `wgpu-core = "${V_WGPU}"`)
+        .replace(/^wgpu-types = .*/m, `wgpu-types = "${V_WGPU}"`)
+        .replace(/^wgpu-hal = .*/m, `wgpu-hal = "${V_WGPU}"`),
+  );
+}
+
+async function main() {
+  await clearTargetDir();
+  await checkoutUpstream();
+  await patchCargo();
+  await bash(join(ROOT_PATH, "tools", "format.js"));
+}
+
+await main();