Webgpu

WebGPU Skill

Use this skill to design, implement, and debug WebGPU applications and GPU compute pipelines. Keep it framework-agnostic and focus on reusable WebGPU/WGSL patterns.

What this skill covers

• Cover WebGPU initialization, device setup, and surface configuration.
• Cover compute pipelines, workgroup sizing, and storage buffer layout.
• Cover render pipelines, render passes, and post-processing patterns.
• Cover GPU/CPU synchronization and safe readback strategies.
• Cover performance and debugging practices.
• Cover architecture patterns: modular passes, phase-based simulation, and capability handling.
• Cover use cases: rendering, compute, ML training/inference, grid simulations, and systems modeling.

Core principles

• Choose a capability strategy: fallback runtime, reduced mode, or fail fast.
• Avoid full GPU readbacks in hot paths; use localized queries or small readback buffers.
• Structure simulation with phases (state, apply, integrate, constrain, correct) to keep WGSL cohesive.
• Use spatial grids or other spatial indexing for neighbor queries and high particle counts.
• Build modular passes so render and compute stages stay composable and testable.

Workflow

When asked to build a WebGPU feature:

1. Confirm the target platform and WebGPU support expectations. 2. Propose a resource layout (buffers, textures, bind groups) with a simple data model. 3. Sketch the pipeline graph (compute vs render passes) and dependencies. 4. Provide minimal working code and scale up with performance constraints. 5. Choose a capability strategy when WebGPU is unavailable.

Deliverable checklist

• Provide clean WebGPU init and error handling.
• Include a buffer layout with alignment notes (16-byte struct alignment for WGSL).
• Include a pass graph with clear read/write ownership (ping-pong textures if needed).
• Call out readback and when it is safe.
• Provide an optional fallback or reduced mode for critical functionality.

References and assets

• Use REFERENCE.md for a compact WebGPU cheat sheet.
• Use references/ for deeper patterns and concepts.
• Use examples/ for runnable snippets.
• Use templates/ for project scaffolds or starter code.

Quick reference

See REFERENCE.md for a compact WebGPU cheat sheet and references/ for deeper patterns, including references/use-cases.md and references/simulation-patterns.md.

WebGPU Skill

This repository contains a WebGPU skill. Skills are reusable capabilities for AI agents, packaged so they can be installed and reused across different tools and workflows. This one is framework-agnostic and focuses on reusable WebGPU/WGSL patterns, orchestration, and performance guidance across a wide range of GPU workloads.

Supported agents

This skill is designed to work with the major agents that support skills, including Claude Code, Codex, Cursor, Cline, and GitHub Copilot, plus other compatible agents listed by the skills ecosystem.

Contents

• SKILL.md: Skill metadata and overview
• REFERENCE.md: Quick reference for core WebGPU patterns
• references/: Concept guides and patterns
• examples/: Small runnable snippets
• templates/: Starter templates

Highlights

• Compute + render pipeline patterns
• Orchestration and phase-based simulation guidance
• Practical performance notes and debugging tips
• Uniform packing and resource layout advice

Installation

Install with the CLI:

npx skills add cazala/webgpu-skill

Refer to SKILL.md for the full entry point.

WebGPU Quick Reference

Device + context

const adapter = await navigator.gpu?.requestAdapter();
if (!adapter) throw new Error("WebGPU not supported");
const device = await adapter.requestDevice();
const context = canvas.getContext("webgpu");
const format = navigator.gpu.getPreferredCanvasFormat();
context.configure({ device, format, alphaMode: "premultiplied" });

Buffer creation

const buffer = device.createBuffer({
  size: byteLength,
  usage: GPUBufferUsage.STORAGE | GPUBufferUsage.COPY_DST,
  mappedAtCreation: false,
});

Uniform packing

When you have many small scalar uniforms, consider packing them into vec4 slots:

struct Params {
  v0: vec4<f32>, // x, y, z, w
  v1: vec4<f32>, // a, b, c, d
};

@group(0) @binding(0) var<uniform> params: Params;

// v0 = [x, y, z, w], v1 = [a, b, c, d]
const u = new Float32Array(8);
u[0] = x;
u[1] = y;
u[2] = z;
u[3] = w;
u[4] = a;
u[5] = b;
u[6] = c;
u[7] = d;
device.queue.writeBuffer(uniformBuffer, 0, u);

This reduces bind group bindings and keeps alignment predictable.

Pipeline setup

const module = device.createShaderModule({ code: wgslSource });
const pipeline = device.createComputePipeline({
  layout: "auto",
  compute: { module, entryPoint: "main" },
});

Dispatch

const encoder = device.createCommandEncoder();
const pass = encoder.beginComputePass();
pass.setPipeline(pipeline);
pass.setBindGroup(0, bindGroup);
pass.dispatchWorkgroups(workgroupsX, workgroupsY, workgroupsZ);
pass.end();
device.queue.submit([encoder.finish()]);

Render pass

const view = context.getCurrentTexture().createView();
const pass = encoder.beginRenderPass({
  colorAttachments: [{
    view,
    clearValue: { r: 0, g: 0, b: 0, a: 1 },
    loadOp: "clear",
    storeOp: "store",
  }],
});

Readback (avoid in hot paths)

const readBuffer = device.createBuffer({
  size: byteLength,
  usage: GPUBufferUsage.COPY_DST | GPUBufferUsage.MAP_READ,
});
encoder.copyBufferToBuffer(srcBuffer, 0, readBuffer, 0, byteLength);
device.queue.submit([encoder.finish()]);
await readBuffer.mapAsync(GPUMapMode.READ);
const data = readBuffer.getMappedRange();

Common pitfalls

• Ensure WGSL structs are aligned to 16 bytes.
• Keep bind group layouts stable to avoid pipeline rebuilds.
• Use ping-pong textures/buffers for multi-pass effects.
• Keep readbacks to bounded results (small buffers).

Compute Shaders

Workgroup sizing

Pick a workgroup size that balances occupancy and memory access. Common sizes are 32, 64, or 128. Make it configurable so you can tune based on the workload and device.

@compute @workgroup_size(64)
fn main(@builtin(global_invocation_id) gid: vec3<u32>) {
  let i = gid.x;
  // ...
}

Storage buffer layout

Use struct-of-arrays when you need coalesced access, and array-of-structs when you need per-element cohesion. Avoid mixing unrelated fields in a single struct if it causes heavy bandwidth.

Simulation phases

For iterative systems, break compute into phases:

• state: pre-pass to compute derived state (densities, activations, caches).
• apply: apply updates (forces, gradient steps, rule updates).
• integrate: advance state with time-step or iteration.
• constrain: enforce bounds, invariants, or stability.
• correct: post-pass fixes, clamping, or projection.

This keeps WGSL readable and makes it easier to insert extra passes.

Neighborhood queries

If you need local context, maintain a spatial grid or tiled buffers on the GPU. This avoids O(n^2) scans and is scalable for large grids or particle counts.

Core Concepts

WebGPU execution model

WebGPU work happens on the GPU via command buffers. The usual flow:

1. Create GPU resources (buffers, textures, samplers). 2. Build pipelines (compute/render). 3. Encode commands into a command encoder. 4. Submit to the device queue.

Resource layout basics

• Uniform buffers: small, frequently updated values.
• Storage buffers: large arrays for compute and general data.
• Textures: render targets or sampled data.

WGSL alignment rules matter. Most structs should align to 16 bytes. If you pack data tightly, confirm offsets and padding.

Bind groups

Bind groups connect buffers/textures to shaders. Keep them stable:

• Create stable bind group layouts.
• Reuse bind groups across frames when possible.
• Use dynamic offsets if you need small per-object variations.

Command encoding

Encode compute passes and render passes. The order you encode is the order the GPU executes. Split passes when you need to:

• Ping-pong textures or buffers
• Read from and write to different resources
• Insert readbacks or copy commands

Debugging and Performance

Avoid heavy readbacks

GPU readbacks stall the pipeline. Prefer:

• Small readback buffers with bounded results
• Localized queries (e.g., "particles within radius")
• Debug-only readbacks behind a toggle

Reduce pipeline churn

• Keep bind group layouts stable
• Avoid rebuilding pipelines per frame
• Batch updates into a single buffer map/write

Frame stability

Clamp delta time to avoid unstable physics. This is especially important when the tab is backgrounded or the GPU is busy.

Validation and error scopes

Use device.pushErrorScope("validation") and device.popErrorScope() during development to catch errors early.

Tuning knobs

Expose and document the key performance knobs:

• workgroup size
• max particles processed per frame
• spatial grid cell size
• neighbor count limit

Rendering Pipelines

Render passes

Render passes write to the swapchain or intermediate textures. Use intermediate textures for post-processing and trails:

• Pass 1: render scene to texture A
• Pass 2: compute/blur into texture B
• Pass 3: composite into swapchain

Ping-pong textures

For iterative effects (trails, blur, simulations), alternate read/write textures each frame. This avoids read-write hazards in the same texture.

Instancing

For large numbers of similar objects (particles, lines), use instancing and keep instance data in a storage buffer.

Color + depth

If you do 3D rendering, add a depth attachment. For 2D systems, skip depth to reduce overhead.

Runtime Fallback Strategy

Why it matters

WebGPU is not guaranteed on every device. Some projects can offer a fallback, while others may choose to fail fast or provide a reduced feature set.

Options

Pick the approach that matches your product goals:

1. Fallback runtime: switch to CPU or a simplified WebGL path. 2. Reduced mode: keep WebGPU-only features gated and disable heavy paths. 3. Fail fast: show a clear error and prompt for a supported device/browser.

Practical tips

• Keep CPU and GPU codepaths in sync for scale and units if you provide fallback.
• Document feature availability for modules that do not support all runtimes.
• Add a feature probe in UI to display the active mode.

Simulation and Orchestration Patterns

These patterns apply to a wide range of WebGPU workloads, from particle systems to neural nets and grid simulations.

Capability handling

Choose the strategy that fits your product:

• Fallback runtime when correctness matters more than speed.
• Reduced mode when only part of the pipeline needs WebGPU.
• Fail fast when the app is WebGPU-only and must guarantee performance.

Phase-based pipelines

Split work into phases that map to compute passes:

1. state: compute derived state (densities, activations, caches). 2. apply: apply updates (forces, gradient steps, rule updates). 3. integrate: advance state with time-step or iteration. 4. constrain: enforce bounds or stability. 5. correct: post-pass fixes or projections.

This keeps shader composition manageable and makes it easier to swap or insert passes.

Spatial or tiled data access

Maintain a spatial index or tiled buffers:

• GPU: bin elements into tiles or buckets to improve locality.
• CPU fallback: use spatial grids or hashed buckets.
• Expose tuning knobs like tileSize or maxNeighbors for performance control.

Localized queries instead of readbacks

Avoid full buffer readbacks. Use a small compute pass to compact a subset into a small buffer and read back only that subset.

Modular passes

Each module or stage declares:

• Its inputs (uniforms, storage buffers, textures).
• Which phase or render pass it contributes to.

The orchestrator composes a final pipeline from these descriptors.

WebGPU Use Cases

WebGPU is useful for a broad range of workloads that benefit from parallel execution:

• Neural networks: inference and training, including backprop and gradient descent.
• Grid simulations: Game of Life, cellular automata, reaction-diffusion.
• Physics and systems: particle dynamics, fluid-like solvers, constraint systems.
• Chaos and dynamics: coupled oscillators, pendulum systems, attractors.
• Signal processing: filters, FFT-based pipelines, image transforms.
• Rendering: forward/deferred rendering, post-processing, UI compositing.

Picking the right model

• Use compute for large parallel workloads with heavy data movement.
• Use render pipelines when rasterization is the core step.
• Combine both when you need compute pre-processing and render post-processing.

WebGPU Project Template

Files

• index.html: canvas and script entry
• main.ts: WebGPU init, compute update, and render loop
• shaders.wgsl: compute + render WGSL

index.html

<!-- Minimal canvas + module entrypoint -->
<canvas id="c"></canvas>
<script type="module" src="/src/main.ts"></script>

main.ts

// Grab the canvas element.
const canvas = document.getElementById("c");
if (!(canvas instanceof HTMLCanvasElement)) {
  throw new Error("Missing canvas");
}

// Request a WebGPU adapter and device.
const adapter = await navigator.gpu?.requestAdapter();
if (!adapter) throw new Error("WebGPU not supported");
const device = await adapter.requestDevice();

// Acquire the WebGPU context from the canvas.
const context = canvas.getContext("webgpu");
if (!context) throw new Error("Missing WebGPU context");

// Simulation parameters.
const particleCount = 4096;
const stride = 4 * 4; // vec2 position + vec2 velocity (4 floats)
const bufferSize = particleCount * stride;

// Initialize particle data in clip-space coordinates.
const particles = new Float32Array(particleCount * 4);
for (let i = 0; i < particleCount; i++) {
  const o = i * 4;
  particles[o + 0] = (Math.random() * 2 - 1) * 0.9; // x in clip space
  particles[o + 1] = (Math.random() * 2 - 1) * 0.9; // y in clip space
  particles[o + 2] = (Math.random() * 2 - 1) * 0.002; // vx
  particles[o + 3] = (Math.random() * 2 - 1) * 0.002; // vy
}

// Create a storage buffer for particle data and upload initial state.
const particleBuffer = device.createBuffer({
  size: bufferSize,
  usage:
    GPUBufferUsage.STORAGE |
    GPUBufferUsage.VERTEX |
    GPUBufferUsage.COPY_DST,
});
device.queue.writeBuffer(particleBuffer, 0, particles);

// Load WGSL source and compile it into a shader module.
const shader = await fetch("/src/shaders.wgsl").then((r) => r.text());
const module = device.createShaderModule({ code: shader });

// Build the compute pipeline for updating particle positions.
const computePipeline = device.createComputePipeline({
  layout: "auto",
  compute: { module, entryPoint: "cs_main" },
});

// Build the render pipeline to draw particles as points.
const renderPipeline = device.createRenderPipeline({
  layout: "auto",
  vertex: { module, entryPoint: "vs_main" },
  fragment: {
    module,
    entryPoint: "fs_main",
    targets: [{ format: navigator.gpu.getPreferredCanvasFormat() }],
  },
  primitive: { topology: "point-list" },
});

// Bind group for compute (storage buffer only).
const computeBindGroup = device.createBindGroup({
  layout: computePipeline.getBindGroupLayout(0),
  entries: [{ binding: 0, resource: { buffer: particleBuffer } }],
});

// Bind group for render (storage buffer read in vertex shader).
const renderBindGroup = device.createBindGroup({
  layout: renderPipeline.getBindGroupLayout(0),
  entries: [{ binding: 0, resource: { buffer: particleBuffer } }],
});

// Configure the canvas context for the preferred swapchain format.
const format = navigator.gpu.getPreferredCanvasFormat();
context.configure({ device, format, alphaMode: "premultiplied" });

function frame() {
  // Keep the canvas resolution in sync with CSS size and device pixel ratio.
  const width = canvas.clientWidth * devicePixelRatio;
  const height = canvas.clientHeight * devicePixelRatio;
  if (canvas.width !== width || canvas.height !== height) {
    canvas.width = width;
    canvas.height = height;
  }

  // Begin a command encoder for this frame.
  const encoder = device.createCommandEncoder();

  // --- Compute pass: update particle positions ---
  const computePass = encoder.beginComputePass();
  computePass.setPipeline(computePipeline);
  computePass.setBindGroup(0, computeBindGroup);
  computePass.dispatchWorkgroups(Math.ceil(particleCount / 256));
  computePass.end();

  // --- Render pass: draw particles as white points ---
  const pass = encoder.beginRenderPass({
    colorAttachments: [{
      view: context.getCurrentTexture().createView(),
      clearValue: { r: 0, g: 0, b: 0, a: 1 },
      loadOp: "clear",
      storeOp: "store",
    }],
  });
  pass.setPipeline(renderPipeline);
  pass.setBindGroup(0, renderBindGroup);
  pass.draw(particleCount);
  pass.end();

  // Submit the frame to the GPU and queue the next tick.
  device.queue.submit([encoder.finish()]);
  requestAnimationFrame(frame);
}

// Kick off the render loop.
frame();

shaders.wgsl

// Particle data layout in the storage buffer.
struct Particle {
  pos: vec2<f32>,
  vel: vec2<f32>,
};

// Storage buffer containing all particles.
@group(0) @binding(0) var<storage, read_write> particles: array<Particle>;

// Compute pass: integrate velocity into position and wrap in clip space.
@compute @workgroup_size(256)
fn cs_main(@builtin(global_invocation_id) gid: vec3<u32>) {
  let i = gid.x;
  if (i >= arrayLength(&particles)) { return; }

  var p = particles[i];
  p.pos += p.vel;

  // Simple wrap-around in clip space to keep particles on screen.
  if (p.pos.x > 1.0) { p.pos.x = -1.0; }
  if (p.pos.x < -1.0) { p.pos.x = 1.0; }
  if (p.pos.y > 1.0) { p.pos.y = -1.0; }
  if (p.pos.y < -1.0) { p.pos.y = 1.0; }

  particles[i] = p;
}

// Vertex shader output (position only).
struct VSOut {
  @builtin(position) position: vec4<f32>,
};

// Vertex pass: read particle position and output clip-space position.
@vertex
fn vs_main(@builtin(vertex_index) vid: u32) -> VSOut {
  let p = particles[vid];
  var out: VSOut;
  out.position = vec4<f32>(p.pos, 0.0, 1.0);
  return out;
}

// Fragment pass: paint every point white.
@fragment
fn fs_main(in: VSOut) -> @location(0) vec4<f32> {
  return vec4<f32>(1.0, 1.0, 1.0, 1.0);
}