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>

18 files changed, 1675 insertions, 10 deletions

diff --git a/cli/args/flags.rs b/cli/args/flags.rs
index 9d7e513f7..8b4fa445e 100644
--- a/cli/args/flags.rs
+++ b/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.
diff --git a/cli/build.rs b/cli/build.rs
index d7f373b99..4adeba944 100644
--- a/cli/build.rs
+++ b/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")
+}
diff --git a/cli/js/40_testing.js b/cli/js/40_testing.js
index 5b51ac169..5ceff938c 100644
--- a/cli/js/40_testing.js
+++ b/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`"],
diff --git a/cli/main.rs b/cli/main.rs
index c95e5dc37..321f32976 100644
--- a/cli/main.rs
+++ b/cli/main.rs
@@ -323,9 +323,14 @@ pub(crate) static UNSTABLE_GRANULAR_FLAGS: &[(
     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,
   ),
 ];
 
diff --git a/cli/tests/integration/js_unit_tests.rs b/cli/tests/integration/js_unit_tests.rs
index 165ab25bf..10bd137d9 100644
--- a/cli/tests/integration/js_unit_tests.rs
+++ b/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,
diff --git a/cli/tests/integration/run_tests.rs b/cli/tests/integration/run_tests.rs
index c96d68b93..32df04483 100644
--- a/cli/tests/integration/run_tests.rs
+++ b/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",
diff --git a/cli/tests/integration/shared_library_tests.rs b/cli/tests/integration/shared_library_tests.rs
index 506c537f6..3e05f8efc 100644
--- a/cli/tests/integration/shared_library_tests.rs
+++ b/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] = [
-    "/System/Library/Frameworks/CoreFoundation.framework/Versions/A/CoreFoundation",
-    "/System/Library/Frameworks/CoreServices.framework/Versions/A/CoreServices",
-    "/usr/lib/libiconv.2.dylib",
-    "/usr/lib/libSystem.B.dylib",
-    "/usr/lib/libobjc.A.dylib",
+  // 	/usr/lib/libobjc.A.dylib (compatibility version 1.0.0, current version 228.0.0)
+
+  // path and whether its weak or not
+  const EXPECTED: [(&str, bool); 9] = [
+    ("/System/Library/Frameworks/CoreFoundation.framework/Versions/A/CoreFoundation", false),
+    ("/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
     );
diff --git a/cli/tests/testdata/run/unstable_webgpu.disabled.out b/cli/tests/testdata/run/unstable_webgpu.disabled.out
new file mode 100644
index 000000000..775866352
--- /dev/null
+++ b/cli/tests/testdata/run/unstable_webgpu.disabled.out
@@ -0,0 +1,2 @@
+main undefined
+worker undefined
diff --git a/cli/tests/testdata/run/unstable_webgpu.enabled.out b/cli/tests/testdata/run/unstable_webgpu.enabled.out
new file mode 100644
index 000000000..e2cc915ba
--- /dev/null
+++ b/cli/tests/testdata/run/unstable_webgpu.enabled.out
@@ -0,0 +1,2 @@
+main [class GPU]
+worker [class GPU]
diff --git a/cli/tests/testdata/run/unstable_webgpu.js b/cli/tests/testdata/run/unstable_webgpu.js
new file mode 100644
index 000000000..a796b1c4d
--- /dev/null
+++ b/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);
+}
diff --git a/cli/tests/testdata/webgpu/computepass_shader.wgsl b/cli/tests/testdata/webgpu/computepass_shader.wgsl
new file mode 100644
index 000000000..41af4363a
--- /dev/null
+++ b/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]);
+}
diff --git a/cli/tests/testdata/webgpu/hellotriangle.out b/cli/tests/testdata/webgpu/hellotriangle.out
new file mode 100644
index 000000000..52972ec9e
--- /dev/null
+++ b/cli/tests/testdata/webgpu/hellotriangle.out
diff --git a/cli/tests/testdata/webgpu/hellotriangle_shader.wgsl b/cli/tests/testdata/webgpu/hellotriangle_shader.wgsl
new file mode 100644
index 000000000..f84ccfe94
--- /dev/null
+++ b/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);
+}
diff --git a/cli/tests/unit/webgpu_test.ts b/cli/tests/unit/webgpu_test.ts
new file mode 100644
index 000000000..2d98167cf
--- /dev/null
+++ b/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;
+    }
+  }
+}
diff --git a/cli/tsc/dts/lib.deno.window.d.ts b/cli/tsc/dts/lib.deno.window.d.ts
index 58b57e52c..0e6acf1e7 100644
--- a/cli/tsc/dts/lib.deno.window.d.ts
+++ b/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;
diff --git a/cli/tsc/dts/lib.deno.worker.d.ts b/cli/tsc/dts/lib.deno.worker.d.ts
index b165f2086..1d95dd483 100644
--- a/cli/tsc/dts/lib.deno.worker.d.ts
+++ b/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;
diff --git a/cli/tsc/dts/lib.deno_webgpu.d.ts b/cli/tsc/dts/lib.deno_webgpu.d.ts
new file mode 100644
index 000000000..9545fdc9e
--- /dev/null
+++ b/cli/tsc/dts/lib.deno_webgpu.d.ts
@@ -0,0 +1,1315 @@
+// Copyright 2018-2023 the Deno authors. All rights reserved. MIT license.
+
+// deno-lint-ignore-file no-explicit-any no-empty-interface
+
+/// <reference no-default-lib="true" />
+/// <reference lib="esnext" />
+
+/** @category WebGPU */
+interface GPUObjectBase {
+  label: string;
+}
+
+/** @category WebGPU */
+declare interface GPUObjectDescriptorBase {
+  label?: string;
+}
+
+/** @category WebGPU */
+declare class GPUSupportedLimits {
+  maxTextureDimension1D?: number;
+  maxTextureDimension2D?: number;
+  maxTextureDimension3D?: number;
+  maxTextureArrayLayers?: number;
+  maxBindGroups?: number;
+  maxBindingsPerBindGroup?: number;
+  maxDynamicUniformBuffersPerPipelineLayout?: number;
+  maxDynamicStorageBuffersPerPipelineLayout?: number;
+  maxSampledTexturesPerShaderStage?: number;
+  maxSamplersPerShaderStage?: number;
+  maxStorageBuffersPerShaderStage?: number;
+  maxStorageTexturesPerShaderStage?: number;
+  maxUniformBuffersPerShaderStage?: number;
+  maxUniformBufferBindingSize?: number;
+  maxStorageBufferBindingSize?: number;
+  minUniformBufferOffsetAlignment?: number;
+  minStorageBufferOffsetAlignment?: number;
+  maxVertexBuffers?: number;
+  maxBufferSize?: number;
+  maxVertexAttributes?: number;
+  maxVertexBufferArrayStride?: number;
+  maxInterStageShaderComponents?: number;
+  maxComputeWorkgroupStorageSize?: number;
+  maxComputeInvocationsPerWorkgroup?: number;
+  maxComputeWorkgroupSizeX?: number;
+  maxComputeWorkgroupSizeY?: number;
+  maxComputeWorkgroupSizeZ?: number;
+  maxComputeWorkgroupsPerDimension?: number;
+}
+
+/** @category WebGPU */
+declare class GPUSupportedFeatures {
+  forEach(
+    callbackfn: (
+      value: GPUFeatureName,
+      value2: GPUFeatureName,
+      set: Set<GPUFeatureName>,
+    ) => void,
+    thisArg?: any,
+  ): void;
+  has(value: GPUFeatureName): boolean;
+  size: number;
+  [Symbol.iterator](): IterableIterator<GPUFeatureName>;
+  entries(): IterableIterator<[GPUFeatureName, GPUFeatureName]>;
+  keys(): IterableIterator<GPUFeatureName>;
+  values(): IterableIterator<GPUFeatureName>;
+}
+
+/** @category WebGPU */
+declare class GPUAdapterInfo {
+  readonly vendor: string;
+  readonly architecture: string;
+  readonly device: string;
+  readonly description: string;
+}
+
+/** @category WebGPU */
+declare class GPU {
+  requestAdapter(
+    options?: GPURequestAdapterOptions,
+  ): Promise<GPUAdapter | null>;
+}
+
+/** @category WebGPU */
+declare interface GPURequestAdapterOptions {
+  powerPreference?: GPUPowerPreference;
+  forceFallbackAdapter?: boolean;
+}
+
+/** @category WebGPU */
+declare type GPUPowerPreference = "low-power" | "high-performance";
+
+/** @category WebGPU */
+declare class GPUAdapter {
+  readonly features: GPUSupportedFeatures;
+  readonly limits: GPUSupportedLimits;
+  readonly isFallbackAdapter: boolean;
+
+  requestDevice(descriptor?: GPUDeviceDescriptor): Promise<GPUDevice>;
+  requestAdapterInfo(unmaskHints?: string[]): Promise<GPUAdapterInfo>;
+}
+
+/** @category WebGPU */
+declare interface GPUDeviceDescriptor extends GPUObjectDescriptorBase {
+  requiredFeatures?: GPUFeatureName[];
+  requiredLimits?: Record<string, number>;
+}
+
+/** @category WebGPU */
+declare type GPUFeatureName =
+  | "depth-clip-control"
+  | "depth32float-stencil8"
+  | "pipeline-statistics-query"
+  | "texture-compression-bc"
+  | "texture-compression-etc2"
+  | "texture-compression-astc"
+  | "timestamp-query"
+  | "indirect-first-instance"
+  | "shader-f16"
+  // extended from spec
+  | "mappable-primary-buffers"
+  | "sampled-texture-binding-array"
+  | "sampled-texture-array-dynamic-indexing"
+  | "sampled-texture-array-non-uniform-indexing"
+  | "unsized-binding-array"
+  | "multi-draw-indirect"
+  | "multi-draw-indirect-count"
+  | "push-constants"
+  | "address-mode-clamp-to-border"
+  | "texture-adapter-specific-format-features"
+  | "shader-float64"
+  | "vertex-attribute-64bit";
+
+/** @category WebGPU */
+declare class GPUDevice extends EventTarget implements GPUObjectBase {
+  label: string;
+
+  readonly lost: Promise<GPUDeviceLostInfo>;
+  pushErrorScope(filter: GPUErrorFilter): undefined;
+  popErrorScope(): Promise<GPUError | null>;
+
+  readonly features: GPUSupportedFeatures;
+  readonly limits: GPUSupportedLimits;
+  readonly queue: GPUQueue;
+
+  destroy(): undefined;
+
+  createBuffer(descriptor: GPUBufferDescriptor): GPUBuffer;
+  createTexture(descriptor: GPUTextureDescriptor): GPUTexture;
+  createSampler(descriptor?: GPUSamplerDescriptor): GPUSampler;
+
+  createBindGroupLayout(
+    descriptor: GPUBindGroupLayoutDescriptor,
+  ): GPUBindGroupLayout;
+  createPipelineLayout(
+    descriptor: GPUPipelineLayoutDescriptor,
+  ): GPUPipelineLayout;
+  createBindGroup(descriptor: GPUBindGroupDescriptor): GPUBindGroup;
+
+  createShaderModule(descriptor: GPUShaderModuleDescriptor): GPUShaderModule;
+  createComputePipeline(
+    descriptor: GPUComputePipelineDescriptor,
+  ): GPUComputePipeline;
+  createRenderPipeline(
+    descriptor: GPURenderPipelineDescriptor,
+  ): GPURenderPipeline;
+  createComputePipelineAsync(
+    descriptor: GPUComputePipelineDescriptor,
+  ): Promise<GPUComputePipeline>;
+  createRenderPipelineAsync(
+    descriptor: GPURenderPipelineDescriptor,
+  ): Promise<GPURenderPipeline>;
+
+  createCommandEncoder(
+    descriptor?: GPUCommandEncoderDescriptor,
+  ): GPUCommandEncoder;
+  createRenderBundleEncoder(
+    descriptor: GPURenderBundleEncoderDescriptor,
+  ): GPURenderBundleEncoder;
+
+  createQuerySet(descriptor: GPUQuerySetDescriptor): GPUQuerySet;
+}
+
+/** @category WebGPU */
+declare class GPUBuffer implements GPUObjectBase {
+  label: string;
+
+  readonly size: number;
+  readonly usage: GPUFlagsConstant;
+  readonly mapState: GPUBufferMapState;
+
+  mapAsync(
+    mode: GPUMapModeFlags,
+    offset?: number,
+    size?: number,
+  ): Promise<undefined>;
+  getMappedRange(offset?: number, size?: number): ArrayBuffer;
+  unmap(): undefined;
+
+  destroy(): undefined;
+}
+
+/** @category WebGPU */
+declare type GPUBufferMapState = "unmapped" | "pending" | "mapped";
+
+/** @category WebGPU */
+declare interface GPUBufferDescriptor extends GPUObjectDescriptorBase {
+  size: number;
+  usage: GPUBufferUsageFlags;
+  mappedAtCreation?: boolean;
+}
+
+/** @category WebGPU */
+declare type GPUBufferUsageFlags = number;
+
+/** @category WebGPU */
+declare type GPUFlagsConstant = number;
+
+/** @category WebGPU */
+declare class GPUBufferUsage {
+  static MAP_READ: 0x0001;
+  static MAP_WRITE: 0x0002;
+  static COPY_SRC: 0x0004;
+  static COPY_DST: 0x0008;
+  static INDEX: 0x0010;
+  static VERTEX: 0x0020;
+  static UNIFORM: 0x0040;
+  static STORAGE: 0x0080;
+  static INDIRECT: 0x0100;
+  static QUERY_RESOLVE: 0x0200;
+}
+
+/** @category WebGPU */
+declare type GPUMapModeFlags = number;
+
+/** @category WebGPU */
+declare class GPUMapMode {
+  static READ: 0x0001;
+  static WRITE: 0x0002;
+}
+
+/** @category WebGPU */
+declare class GPUTexture implements GPUObjectBase {
+  label: string;
+
+  createView(descriptor?: GPUTextureViewDescriptor): GPUTextureView;
+  destroy(): undefined;
+
+  readonly width: number;
+  readonly height: number;
+  readonly depthOrArrayLayers: number;
+  readonly mipLevelCount: number;
+  readonly sampleCount: number;
+  readonly dimension: GPUTextureDimension;
+  readonly format: GPUTextureFormat;
+  readonly usage: GPUFlagsConstant;
+}
+
+/** @category WebGPU */
+declare interface GPUTextureDescriptor extends GPUObjectDescriptorBase {
+  size: GPUExtent3D;
+  mipLevelCount?: number;
+  sampleCount?: number;
+  dimension?: GPUTextureDimension;
+  format: GPUTextureFormat;
+  usage: GPUTextureUsageFlags;
+  viewFormats?: GPUTextureFormat[];
+}
+
+/** @category WebGPU */
+declare type GPUTextureDimension = "1d" | "2d" | "3d";
+
+/** @category WebGPU */
+declare type GPUTextureUsageFlags = number;
+
+/** @category WebGPU */
+declare class GPUTextureUsage {
+  static COPY_SRC: 0x01;
+  static COPY_DST: 0x02;
+  static TEXTURE_BINDING: 0x04;
+  static STORAGE_BINDING: 0x08;
+  static RENDER_ATTACHMENT: 0x10;
+}
+
+/** @category WebGPU */
+declare class GPUTextureView implements GPUObjectBase {
+  label: string;
+}
+
+/** @category WebGPU */
+declare interface GPUTextureViewDescriptor extends GPUObjectDescriptorBase {
+  format?: GPUTextureFormat;
+  dimension?: GPUTextureViewDimension;
+  aspect?: GPUTextureAspect;
+  baseMipLevel?: number;
+  mipLevelCount?: number;
+  baseArrayLayer?: number;
+  arrayLayerCount?: number;
+}
+
+/** @category WebGPU */
+declare type GPUTextureViewDimension =
+  | "1d"
+  | "2d"
+  | "2d-array"
+  | "cube"
+  | "cube-array"
+  | "3d";
+
+/** @category WebGPU */
+declare type GPUTextureAspect = "all" | "stencil-only" | "depth-only";
+
+/** @category WebGPU */
+declare type GPUTextureFormat =
+  | "r8unorm"
+  | "r8snorm"
+  | "r8uint"
+  | "r8sint"
+  | "r16uint"
+  | "r16sint"
+  | "r16float"
+  | "rg8unorm"
+  | "rg8snorm"
+  | "rg8uint"
+  | "rg8sint"
+  | "r32uint"
+  | "r32sint"
+  | "r32float"
+  | "rg16uint"
+  | "rg16sint"
+  | "rg16float"
+  | "rgba8unorm"
+  | "rgba8unorm-srgb"
+  | "rgba8snorm"
+  | "rgba8uint"
+  | "rgba8sint"
+  | "bgra8unorm"
+  | "bgra8unorm-srgb"
+  | "rgb9e5ufloat"
+  | "rgb10a2unorm"
+  | "rg11b10ufloat"
+  | "rg32uint"
+  | "rg32sint"
+  | "rg32float"
+  | "rgba16uint"
+  | "rgba16sint"
+  | "rgba16float"
+  | "rgba32uint"
+  | "rgba32sint"
+  | "rgba32float"
+  | "stencil8"
+  | "depth16unorm"
+  | "depth24plus"
+  | "depth24plus-stencil8"
+  | "depth32float"
+  | "depth32float-stencil8"
+  | "bc1-rgba-unorm"
+  | "bc1-rgba-unorm-srgb"
+  | "bc2-rgba-unorm"
+  | "bc2-rgba-unorm-srgb"
+  | "bc3-rgba-unorm"
+  | "bc3-rgba-unorm-srgb"
+  | "bc4-r-unorm"
+  | "bc4-r-snorm"
+  | "bc5-rg-unorm"
+  | "bc5-rg-snorm"
+  | "bc6h-rgb-ufloat"
+  | "bc6h-rgb-float"
+  | "bc7-rgba-unorm"
+  | "bc7-rgba-unorm-srgb"
+  | "etc2-rgb8unorm"
+  | "etc2-rgb8unorm-srgb"
+  | "etc2-rgb8a1unorm"
+  | "etc2-rgb8a1unorm-srgb"
+  | "etc2-rgba8unorm"
+  | "etc2-rgba8unorm-srgb"
+  | "eac-r11unorm"
+  | "eac-r11snorm"
+  | "eac-rg11unorm"
+  | "eac-rg11snorm"
+  | "astc-4x4-unorm"
+  | "astc-4x4-unorm-srgb"
+  | "astc-5x4-unorm"
+  | "astc-5x4-unorm-srgb"
+  | "astc-5x5-unorm"
+  | "astc-5x5-unorm-srgb"
+  | "astc-6x5-unorm"
+  | "astc-6x5-unorm-srgb"
+  | "astc-6x6-unorm"
+  | "astc-6x6-unorm-srgb"
+  | "astc-8x5-unorm"
+  | "astc-8x5-unorm-srgb"
+  | "astc-8x6-unorm"
+  | "astc-8x6-unorm-srgb"
+  | "astc-8x8-unorm"
+  | "astc-8x8-unorm-srgb"
+  | "astc-10x5-unorm"
+  | "astc-10x5-unorm-srgb"
+  | "astc-10x6-unorm"
+  | "astc-10x6-unorm-srgb"
+  | "astc-10x8-unorm"
+  | "astc-10x8-unorm-srgb"
+  | "astc-10x10-unorm"
+  | "astc-10x10-unorm-srgb"
+  | "astc-12x10-unorm"
+  | "astc-12x10-unorm-srgb"
+  | "astc-12x12-unorm"
+  | "astc-12x12-unorm-srgb";
+
+/** @category WebGPU */
+declare class GPUSampler implements GPUObjectBase {
+  label: string;
+}
+
+/** @category WebGPU */
+declare interface GPUSamplerDescriptor extends GPUObjectDescriptorBase {
+  addressModeU?: GPUAddressMode;
+  addressModeV?: GPUAddressMode;
+  addressModeW?: GPUAddressMode;
+  magFilter?: GPUFilterMode;
+  minFilter?: GPUFilterMode;
+  mipmapFilter?: GPUMipmapFilterMode;
+  lodMinClamp?: number;
+  lodMaxClamp?: number;
+  compare?: GPUCompareFunction;
+  maxAnisotropy?: number;
+}
+
+/** @category WebGPU */
+declare type GPUAddressMode = "clamp-to-edge" | "repeat" | "mirror-repeat";
+
+/** @category WebGPU */
+declare type GPUFilterMode = "nearest" | "linear";
+
+/** @category WebGPU */
+declare type GPUMipmapFilterMode = "nearest" | "linear";
+
+/** @category WebGPU */
+declare type GPUCompareFunction =
+  | "never"
+  | "less"
+  | "equal"
+  | "less-equal"
+  | "greater"
+  | "not-equal"
+  | "greater-equal"
+  | "always";
+
+/** @category WebGPU */
+declare class GPUBindGroupLayout implements GPUObjectBase {
+  label: string;
+}
+
+/** @category WebGPU */
+declare interface GPUBindGroupLayoutDescriptor extends GPUObjectDescriptorBase {
+  entries: GPUBindGroupLayoutEntry[];
+}
+
+/** @category WebGPU */
+declare interface GPUBindGroupLayoutEntry {
+  binding: number;
+  visibility: GPUShaderStageFlags;
+
+  buffer?: GPUBufferBindingLayout;
+  sampler?: GPUSamplerBindingLayout;
+  texture?: GPUTextureBindingLayout;
+  storageTexture?: GPUStorageTextureBindingLayout;
+}
+
+/** @category WebGPU */
+declare type GPUShaderStageFlags = number;
+
+/** @category WebGPU */
+declare class GPUShaderStage {
+  static VERTEX: 0x1;
+  static FRAGMENT: 0x2;
+  static COMPUTE: 0x4;
+}
+
+/** @category WebGPU */
+declare interface GPUBufferBindingLayout {
+  type?: GPUBufferBindingType;
+  hasDynamicOffset?: boolean;
+  minBindingSize?: number;
+}
+
+/** @category WebGPU */
+declare type GPUBufferBindingType = "uniform" | "storage" | "read-only-storage";
+
+/** @category WebGPU */
+declare interface GPUSamplerBindingLayout {
+  type?: GPUSamplerBindingType;
+}
+
+/** @category WebGPU */
+declare type GPUSamplerBindingType =
+  | "filtering"
+  | "non-filtering"
+  | "comparison";
+
+/** @category WebGPU */
+declare interface GPUTextureBindingLayout {
+  sampleType?: GPUTextureSampleType;
+  viewDimension?: GPUTextureViewDimension;
+  multisampled?: boolean;
+}
+
+/** @category WebGPU */
+declare type GPUTextureSampleType =
+  | "float"
+  | "unfilterable-float"
+  | "depth"
+  | "sint"
+  | "uint";
+
+/** @category WebGPU */
+declare type GPUStorageTextureAccess = "write-only";
+
+/** @category WebGPU */
+declare interface GPUStorageTextureBindingLayout {
+  access: GPUStorageTextureAccess;
+  format: GPUTextureFormat;
+  viewDimension?: GPUTextureViewDimension;
+}
+
+/** @category WebGPU */
+declare class GPUBindGroup implements GPUObjectBase {
+  label: string;
+}
+
+/** @category WebGPU */
+declare interface GPUBindGroupDescriptor extends GPUObjectDescriptorBase {
+  layout: GPUBindGroupLayout;
+  entries: GPUBindGroupEntry[];
+}
+
+/** @category WebGPU */
+declare type GPUBindingResource =
+  | GPUSampler
+  | GPUTextureView
+  | GPUBufferBinding;
+
+/** @category WebGPU */
+declare interface GPUBindGroupEntry {
+  binding: number;
+  resource: GPUBindingResource;
+}
+
+/** @category WebGPU */
+declare interface GPUBufferBinding {
+  buffer: GPUBuffer;
+  offset?: number;
+  size?: number;
+}
+
+/** @category WebGPU */
+declare class GPUPipelineLayout implements GPUObjectBase {
+  label: string;
+}
+
+/** @category WebGPU */
+declare interface GPUPipelineLayoutDescriptor extends GPUObjectDescriptorBase {
+  bindGroupLayouts: GPUBindGroupLayout[];
+}
+
+/** @category WebGPU */
+declare type GPUCompilationMessageType = "error" | "warning" | "info";
+
+/** @category WebGPU */
+declare interface GPUCompilationMessage {
+  readonly message: string;
+  readonly type: GPUCompilationMessageType;
+  readonly lineNum: number;
+  readonly linePos: number;
+}
+
+/** @category WebGPU */
+declare interface GPUCompilationInfo {
+  readonly messages: ReadonlyArray<GPUCompilationMessage>;
+}
+
+/** @category WebGPU */
+declare class GPUShaderModule implements GPUObjectBase {
+  label: string;
+}
+
+/** @category WebGPU */
+declare interface GPUShaderModuleDescriptor extends GPUObjectDescriptorBase {
+  code: string;
+  sourceMap?: any;
+}
+
+/** @category WebGPU */
+declare type GPUAutoLayoutMode = "auto";
+
+/** @category WebGPU */
+declare interface GPUPipelineDescriptorBase extends GPUObjectDescriptorBase {
+  layout: GPUPipelineLayout | GPUAutoLayoutMode;
+}
+
+/** @category WebGPU */
+declare interface GPUPipelineBase {
+  getBindGroupLayout(index: number): GPUBindGroupLayout;
+}
+
+/** @category WebGPU */
+declare interface GPUProgrammableStage {
+  module: GPUShaderModule;
+  entryPoint: string;
+}
+
+/** @category WebGPU */
+declare class GPUComputePipeline implements GPUObjectBase, GPUPipelineBase {
+  label: string;
+
+  getBindGroupLayout(index: number): GPUBindGroupLayout;
+}
+
+/** @category WebGPU */
+declare interface GPUComputePipelineDescriptor
+  extends GPUPipelineDescriptorBase {
+  compute: GPUProgrammableStage;
+}
+
+/** @category WebGPU */
+declare class GPURenderPipeline implements GPUObjectBase, GPUPipelineBase {
+  label: string;
+
+  getBindGroupLayout(index: number): GPUBindGroupLayout;
+}
+
+/** @category WebGPU */
+declare interface GPURenderPipelineDescriptor
+  extends GPUPipelineDescriptorBase {
+  vertex: GPUVertexState;
+  primitive?: GPUPrimitiveState;
+  depthStencil?: GPUDepthStencilState;
+  multisample?: GPUMultisampleState;
+  fragment?: GPUFragmentState;
+}
+
+/** @category WebGPU */
+declare interface GPUPrimitiveState {
+  topology?: GPUPrimitiveTopology;
+  stripIndexFormat?: GPUIndexFormat;
+  frontFace?: GPUFrontFace;
+  cullMode?: GPUCullMode;
+  unclippedDepth?: boolean;
+}
+
+/** @category WebGPU */
+declare type GPUPrimitiveTopology =
+  | "point-list"
+  | "line-list"
+  | "line-strip"
+  | "triangle-list"
+  | "triangle-strip";
+
+/** @category WebGPU */
+declare type GPUFrontFace = "ccw" | "cw";
+
+/** @category WebGPU */
+declare type GPUCullMode = "none" | "front" | "back";
+
+/** @category WebGPU */
+declare interface GPUMultisampleState {
+  count?: number;
+  mask?: number;
+  alphaToCoverageEnabled?: boolean;
+}
+
+/** @category WebGPU */
+declare interface GPUFragmentState extends GPUProgrammableStage {
+  targets: (GPUColorTargetState | null)[];
+}
+
+/** @category WebGPU */
+declare interface GPUColorTargetState {
+  format: GPUTextureFormat;
+
+  blend?: GPUBlendState;
+  writeMask?: GPUColorWriteFlags;
+}
+
+/** @category WebGPU */
+declare interface GPUBlendState {
+  color: GPUBlendComponent;
+  alpha: GPUBlendComponent;
+}
+
+/** @category WebGPU */
+declare type GPUColorWriteFlags = number;
+
+/** @category WebGPU */
+declare class GPUColorWrite {
+  static RED: 0x1;
+  static GREEN: 0x2;
+  static BLUE: 0x4;
+  static ALPHA: 0x8;
+  static ALL: 0xF;
+}
+
+/** @category WebGPU */
+declare interface GPUBlendComponent {
+  operation?: GPUBlendOperation;
+  srcFactor?: GPUBlendFactor;
+  dstFactor?: GPUBlendFactor;
+}
+
+/** @category WebGPU */
+declare type GPUBlendFactor =
+  | "zero"
+  | "one"
+  | "src"
+  | "one-minus-src"
+  | "src-alpha"
+  | "one-minus-src-alpha"
+  | "dst"
+  | "one-minus-dst"
+  | "dst-alpha"
+  | "one-minus-dst-alpha"
+  | "src-alpha-saturated"
+  | "constant"
+  | "one-minus-constant";
+
+/** @category WebGPU */
+declare type GPUBlendOperation =
+  | "add"
+  | "subtract"
+  | "reverse-subtract"
+  | "min"
+  | "max";
+
+/** @category WebGPU */
+declare interface GPUDepthStencilState {
+  format: GPUTextureFormat;
+
+  depthWriteEnabled: boolean;
+  depthCompare: GPUCompareFunction;
+
+  stencilFront?: GPUStencilFaceState;
+  stencilBack?: GPUStencilFaceState;
+
+  stencilReadMask?: number;
+  stencilWriteMask?: number;
+
+  depthBias?: number;
+  depthBiasSlopeScale?: number;
+  depthBiasClamp?: number;
+}
+
+/** @category WebGPU */
+declare interface GPUStencilFaceState {
+  compare?: GPUCompareFunction;
+  failOp?: GPUStencilOperation;
+  depthFailOp?: GPUStencilOperation;
+  passOp?: GPUStencilOperation;
+}
+
+/** @category WebGPU */
+declare type GPUStencilOperation =
+  | "keep"
+  | "zero"
+  | "replace"
+  | "invert"
+  | "increment-clamp"
+  | "decrement-clamp"
+  | "increment-wrap"
+  | "decrement-wrap";
+
+/** @category WebGPU */
+declare type GPUIndexFormat = "uint16" | "uint32";
+
+/** @category WebGPU */
+declare type GPUVertexFormat =
+  | "uint8x2"
+  | "uint8x4"
+  | "sint8x2"
+  | "sint8x4"
+  | "unorm8x2"
+  | "unorm8x4"
+  | "snorm8x2"
+  | "snorm8x4"
+  | "uint16x2"
+  | "uint16x4"
+  | "sint16x2"
+  | "sint16x4"
+  | "unorm16x2"
+  | "unorm16x4"
+  | "snorm16x2"
+  | "snorm16x4"
+  | "float16x2"
+  | "float16x4"
+  | "float32"
+  | "float32x2"
+  | "float32x3"
+  | "float32x4"
+  | "uint32"
+  | "uint32x2"
+  | "uint32x3"
+  | "uint32x4"
+  | "sint32"
+  | "sint32x2"
+  | "sint32x3"
+  | "sint32x4";
+
+/** @category WebGPU */
+declare type GPUVertexStepMode = "vertex" | "instance";
+
+/** @category WebGPU */
+declare interface GPUVertexState extends GPUProgrammableStage {
+  buffers?: (GPUVertexBufferLayout | null)[];
+}
+
+/** @category WebGPU */
+declare interface GPUVertexBufferLayout {
+  arrayStride: number;
+  stepMode?: GPUVertexStepMode;
+  attributes: GPUVertexAttribute[];
+}
+
+/** @category WebGPU */
+declare interface GPUVertexAttribute {
+  format: GPUVertexFormat;
+  offset: number;
+
+  shaderLocation: number;
+}
+
+/** @category WebGPU */
+declare interface GPUImageDataLayout {
+  offset?: number;
+  bytesPerRow?: number;
+  rowsPerImage?: number;
+}
+
+/** @category WebGPU */
+declare class GPUCommandBuffer implements GPUObjectBase {
+  label: string;
+}
+
+/** @category WebGPU */
+declare interface GPUCommandBufferDescriptor extends GPUObjectDescriptorBase {}
+
+/** @category WebGPU */
+declare class GPUCommandEncoder implements GPUObjectBase {
+  label: string;
+
+  beginRenderPass(descriptor: GPURenderPassDescriptor): GPURenderPassEncoder;
+  beginComputePass(
+    descriptor?: GPUComputePassDescriptor,
+  ): GPUComputePassEncoder;
+
+  copyBufferToBuffer(
+    source: GPUBuffer,
+    sourceOffset: number,
+    destination: GPUBuffer,
+    destinationOffset: number,
+    size: number,
+  ): undefined;
+
+  copyBufferToTexture(
+    source: GPUImageCopyBuffer,
+    destination: GPUImageCopyTexture,
+    copySize: GPUExtent3D,
+  ): undefined;
+
+  copyTextureToBuffer(
+    source: GPUImageCopyTexture,
+    destination: GPUImageCopyBuffer,
+    copySize: GPUExtent3D,
+  ): undefined;
+
+  copyTextureToTexture(
+    source: GPUImageCopyTexture,
+    destination: GPUImageCopyTexture,
+    copySize: GPUExtent3D,
+  ): undefined;
+
+  clearBuffer(
+    destination: GPUBuffer,
+    destinationOffset?: number,
+    size?: number,
+  ): undefined;
+
+  pushDebugGroup(groupLabel: string): undefined;
+  popDebugGroup(): undefined;
+  insertDebugMarker(markerLabel: string): undefined;
+
+  writeTimestamp(querySet: GPUQuerySet, queryIndex: number): undefined;
+
+  resolveQuerySet(
+    querySet: GPUQuerySet,
+    firstQuery: number,
+    queryCount: number,
+    destination: GPUBuffer,
+    destinationOffset: number,
+  ): undefined;
+
+  finish(descriptor?: GPUCommandBufferDescriptor): GPUCommandBuffer;
+}
+
+/** @category WebGPU */
+declare interface GPUCommandEncoderDescriptor extends GPUObjectDescriptorBase {}
+
+/** @category WebGPU */
+declare interface GPUImageCopyBuffer extends GPUImageDataLayout {
+  buffer: GPUBuffer;
+}
+
+/** @category WebGPU */
+declare interface GPUImageCopyTexture {
+  texture: GPUTexture;
+  mipLevel?: number;
+  origin?: GPUOrigin3D;
+  aspect?: GPUTextureAspect;
+}
+
+/** @category WebGPU */
+interface GPUProgrammablePassEncoder {
+  setBindGroup(
+    index: number,
+    bindGroup: GPUBindGroup,
+    dynamicOffsets?: number[],
+  ): undefined;
+
+  setBindGroup(
+    index: number,
+    bindGroup: GPUBindGroup,
+    dynamicOffsetsData: Uint32Array,
+    dynamicOffsetsDataStart: number,
+    dynamicOffsetsDataLength: number,
+  ): undefined;
+
+  pushDebugGroup(groupLabel: string): undefined;
+  popDebugGroup(): undefined;
+  insertDebugMarker(markerLabel: string): undefined;
+}
+
+/** @category WebGPU */
+declare class GPUComputePassEncoder
+  implements GPUObjectBase, GPUProgrammablePassEncoder {
+  label: string;
+  setBindGroup(
+    index: number,
+    bindGroup: GPUBindGroup,
+    dynamicOffsets?: number[],
+  ): undefined;
+  setBindGroup(
+    index: number,
+    bindGroup: GPUBindGroup,
+    dynamicOffsetsData: Uint32Array,
+    dynamicOffsetsDataStart: number,
+    dynamicOffsetsDataLength: number,
+  ): undefined;
+  pushDebugGroup(groupLabel: string): undefined;
+  popDebugGroup(): undefined;
+  insertDebugMarker(markerLabel: string): undefined;
+  setPipeline(pipeline: GPUComputePipeline): undefined;
+  dispatchWorkgroups(x: number, y?: number, z?: number): undefined;
+  dispatchWorkgroupsIndirect(
+    indirectBuffer: GPUBuffer,
+    indirectOffset: number,
+  ): undefined;
+
+  end(): undefined;
+}
+
+/** @category WebGPU */
+declare interface GPUComputePassTimestampWrites {
+  querySet: GPUQuerySet;
+  beginningOfPassWriteIndex?: number;
+  endOfPassWriteIndex?: number;
+}
+
+/** @category WebGPU */
+declare interface GPUComputePassDescriptor extends GPUObjectDescriptorBase {
+  timestampWrites?: GPUComputePassTimestampWrites;
+}
+
+/** @category WebGPU */
+interface GPURenderEncoderBase {
+  setPipeline(pipeline: GPURenderPipeline): undefined;
+
+  setIndexBuffer(
+    buffer: GPUBuffer,
+    indexFormat: GPUIndexFormat,
+    offset?: number,
+    size?: number,
+  ): undefined;
+  setVertexBuffer(
+    slot: number,
+    buffer: GPUBuffer,
+    offset?: number,
+    size?: number,
+  ): undefined;
+
+  draw(
+    vertexCount: number,
+    instanceCount?: number,
+    firstVertex?: number,
+    firstInstance?: number,
+  ): undefined;
+  drawIndexed(
+    indexCount: number,
+    instanceCount?: number,
+    firstIndex?: number,
+    baseVertex?: number,
+    firstInstance?: number,
+  ): undefined;
+
+  drawIndirect(indirectBuffer: GPUBuffer, indirectOffset: number): undefined;
+  drawIndexedIndirect(
+    indirectBuffer: GPUBuffer,
+    indirectOffset: number,
+  ): undefined;
+}
+
+/** @category WebGPU */
+declare class GPURenderPassEncoder
+  implements GPUObjectBase, GPUProgrammablePassEncoder, GPURenderEncoderBase {
+  label: string;
+  setBindGroup(
+    index: number,
+    bindGroup: GPUBindGroup,
+    dynamicOffsets?: number[],
+  ): undefined;
+  setBindGroup(
+    index: number,
+    bindGroup: GPUBindGroup,
+    dynamicOffsetsData: Uint32Array,
+    dynamicOffsetsDataStart: number,
+    dynamicOffsetsDataLength: number,
+  ): undefined;
+  pushDebugGroup(groupLabel: string): undefined;
+  popDebugGroup(): undefined;
+  insertDebugMarker(markerLabel: string): undefined;
+  setPipeline(pipeline: GPURenderPipeline): undefined;
+  setIndexBuffer(
+    buffer: GPUBuffer,
+    indexFormat: GPUIndexFormat,
+    offset?: number,
+    size?: number,
+  ): undefined;
+  setVertexBuffer(
+    slot: number,
+    buffer: GPUBuffer,
+    offset?: number,
+    size?: number,
+  ): undefined;
+  draw(
+    vertexCount: number,
+    instanceCount?: number,
+    firstVertex?: number,
+    firstInstance?: number,
+  ): undefined;
+  drawIndexed(
+    indexCount: number,
+    instanceCount?: number,
+    firstIndex?: number,
+    baseVertex?: number,
+    firstInstance?: number,
+  ): undefined;
+  drawIndirect(indirectBuffer: GPUBuffer, indirectOffset: number): undefined;
+  drawIndexedIndirect(
+    indirectBuffer: GPUBuffer,
+    indirectOffset: number,
+  ): undefined;
+
+  setViewport(
+    x: number,
+    y: number,
+    width: number,
+    height: number,
+    minDepth: number,
+    maxDepth: number,
+  ): undefined;
+
+  setScissorRect(
+    x: number,
+    y: number,
+    width: number,
+    height: number,
+  ): undefined;
+
+  setBlendConstant(color: GPUColor): undefined;
+  setStencilReference(reference: number): undefined;
+
+  beginOcclusionQuery(queryIndex: number): undefined;
+  endOcclusionQuery(): undefined;
+
+  executeBundles(bundles: GPURenderBundle[]): undefined;
+  end(): undefined;
+}
+
+/** @category WebGPU */
+declare interface GPURenderPassTimestampWrites {
+  querySet: GPUQuerySet;
+  beginningOfPassWriteIndex?: number;
+  endOfPassWriteIndex?: number;
+}
+
+/** @category WebGPU */
+declare interface GPURenderPassDescriptor extends GPUObjectDescriptorBase {
+  colorAttachments: (GPURenderPassColorAttachment | null)[];
+  depthStencilAttachment?: GPURenderPassDepthStencilAttachment;
+  occlusionQuerySet?: GPUQuerySet;
+  timestampWrites?: GPURenderPassTimestampWrites;
+}
+
+/** @category WebGPU */
+declare interface GPURenderPassColorAttachment {
+  view: GPUTextureView;
+  resolveTarget?: GPUTextureView;
+
+  clearValue?: GPUColor;
+  loadOp: GPULoadOp;
+  storeOp: GPUStoreOp;
+}
+
+/** @category WebGPU */
+declare interface GPURenderPassDepthStencilAttachment {
+  view: GPUTextureView;
+
+  depthClearValue?: number;
+  depthLoadOp?: GPULoadOp;
+  depthStoreOp?: GPUStoreOp;
+  depthReadOnly?: boolean;
+
+  stencilClearValue?: number;
+  stencilLoadOp?: GPULoadOp;
+  stencilStoreOp?: GPUStoreOp;
+  stencilReadOnly?: boolean;
+}
+
+/** @category WebGPU */
+declare type GPULoadOp = "load" | "clear";
+
+/** @category WebGPU */
+declare type GPUStoreOp = "store" | "discard";
+
+/** @category WebGPU */
+declare class GPURenderBundle implements GPUObjectBase {
+  label: string;
+}
+
+/** @category WebGPU */
+declare interface GPURenderBundleDescriptor extends GPUObjectDescriptorBase {}
+
+/** @category WebGPU */
+declare class GPURenderBundleEncoder
+  implements GPUObjectBase, GPUProgrammablePassEncoder, GPURenderEncoderBase {
+  label: string;
+  draw(
+    vertexCount: number,
+    instanceCount?: number,
+    firstVertex?: number,
+    firstInstance?: number,
+  ): undefined;
+  drawIndexed(
+    indexCount: number,
+    instanceCount?: number,
+    firstIndex?: number,
+    baseVertex?: number,
+    firstInstance?: number,
+  ): undefined;
+  drawIndexedIndirect(
+    indirectBuffer: GPUBuffer,
+    indirectOffset: number,
+  ): undefined;
+  drawIndirect(indirectBuffer: GPUBuffer, indirectOffset: number): undefined;
+  insertDebugMarker(markerLabel: string): undefined;
+  popDebugGroup(): undefined;
+  pushDebugGroup(groupLabel: string): undefined;
+  setBindGroup(
+    index: number,
+    bindGroup: GPUBindGroup,
+    dynamicOffsets?: number[],
+  ): undefined;
+  setBindGroup(
+    index: number,
+    bindGroup: GPUBindGroup,
+    dynamicOffsetsData: Uint32Array,
+    dynamicOffsetsDataStart: number,
+    dynamicOffsetsDataLength: number,
+  ): undefined;
+  setIndexBuffer(
+    buffer: GPUBuffer,
+    indexFormat: GPUIndexFormat,
+    offset?: number,
+    size?: number,
+  ): undefined;
+  setPipeline(pipeline: GPURenderPipeline): undefined;
+  setVertexBuffer(
+    slot: number,
+    buffer: GPUBuffer,
+    offset?: number,
+    size?: number,
+  ): undefined;
+
+  finish(descriptor?: GPURenderBundleDescriptor): GPURenderBundle;
+}
+
+/** @category WebGPU */
+declare interface GPURenderPassLayout extends GPUObjectDescriptorBase {
+  colorFormats: (GPUTextureFormat | null)[];
+  depthStencilFormat?: GPUTextureFormat;
+  sampleCount?: number;
+}
+
+/** @category WebGPU */
+declare interface GPURenderBundleEncoderDescriptor extends GPURenderPassLayout {
+  depthReadOnly?: boolean;
+  stencilReadOnly?: boolean;
+}
+
+/** @category WebGPU */
+declare class GPUQueue implements GPUObjectBase {
+  label: string;
+
+  submit(commandBuffers: GPUCommandBuffer[]): undefined;
+
+  onSubmittedWorkDone(): Promise<undefined>;
+
+  writeBuffer(
+    buffer: GPUBuffer,
+    bufferOffset: number,
+    data: BufferSource,
+    dataOffset?: number,
+    size?: number,
+  ): undefined;
+
+  writeTexture(
+    destination: GPUImageCopyTexture,
+    data: BufferSource,
+    dataLayout: GPUImageDataLayout,
+    size: GPUExtent3D,
+  ): undefined;
+}
+
+/** @category WebGPU */
+declare class GPUQuerySet implements GPUObjectBase {
+  label: string;
+
+  destroy(): undefined;
+
+  readonly type: GPUQueryType;
+  readonly count: number;
+}
+
+/** @category WebGPU */
+declare interface GPUQuerySetDescriptor extends GPUObjectDescriptorBase {
+  type: GPUQueryType;
+  count: number;
+}
+
+/** @category WebGPU */
+declare type GPUQueryType = "occlusion" | "timestamp";
+
+/** @category WebGPU */
+declare type GPUDeviceLostReason = "destroyed";
+
+/** @category WebGPU */
+declare interface GPUDeviceLostInfo {
+  readonly reason: GPUDeviceLostReason;
+  readonly message: string;
+}
+
+/** @category WebGPU */
+declare class GPUError {
+  readonly message: string;
+}
+
+/** @category WebGPU */
+declare class GPUOutOfMemoryError extends GPUError {
+  constructor(message: string);
+}
+
+/** @category WebGPU */
+declare class GPUValidationError extends GPUError {
+  constructor(message: string);
+}
+
+/** @category WebGPU */
+declare type GPUErrorFilter = "out-of-memory" | "validation";
+
+/** @category WebGPU */
+declare interface GPUColorDict {
+  r: number;
+  g: number;
+  b: number;
+  a: number;
+}
+
+/** @category WebGPU */
+declare type GPUColor = number[] | GPUColorDict;
+
+/** @category WebGPU */
+declare interface GPUOrigin3DDict {
+  x?: number;
+  y?: number;
+  z?: number;
+}
+
+/** @category WebGPU */
+declare type GPUOrigin3D = number[] | GPUOrigin3DDict;
+
+/** @category WebGPU */
+declare interface GPUExtent3DDict {
+  width: number;
+  height?: number;
+  depthOrArrayLayers?: number;
+}
+
+/** @category WebGPU */
+declare type GPUExtent3D = number[] | GPUExtent3DDict;
diff --git a/cli/tsc/mod.rs b/cli/tsc/mod.rs
index 0c3a5602e..2a820a6ee 100644
--- a/cli/tsc/mod.rs
+++ b/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",