WebGPU

export async function initGPU() {
  if (!navigator.gpu) throw new Error("WebGPU unsupported");
  const adapter = await navigator.gpu.requestAdapter({
    powerPreference: "high-performance",
  });
  if (!adapter) throw new Error("No suitable GPU adapter");
 
  const hasF16 = adapter.features.has("shader-f16");
  const device = await adapter.requestDevice({
    label: "main",
    requiredFeatures: hasF16 ? ["shader-f16"] : [],
    requiredLimits: {
      // ask for more than the default only when needed
      maxStorageBufferBindingSize:
        adapter.limits.maxStorageBufferBindingSize,
    },
  });
  return { adapter, device, hasF16 };
}

export function configureCanvas(
  canvas: HTMLCanvasElement, device: GPUDevice,
) {
  const context = canvas.getContext("webgpu");
  if (!context) throw new Error("webgpu context failed");
  const format = navigator.gpu.getPreferredCanvasFormat();
  context.configure({
    device,
    format,              // "bgra8unorm" or "rgba8unorm"
    alphaMode: "premultiplied", // default "opaque"
  });
  return { context, format };
}

struct Uniforms {
  mvp: mat4x4f,
  tint: vec4f,
};
 
@group(0) @binding(0) var<uniform> u: Uniforms;
@group(0) @binding(1) var tex: texture_2d<f32>;
@group(0) @binding(2) var samp: sampler;
 
struct VSOut {
  @builtin(position) pos: vec4f,
  @location(0) uv: vec2f,
};
 
@vertex
fn vs(
  @location(0) pos: vec3f,
  @location(1) uv: vec2f,
) -> VSOut {
  var out: VSOut;
  out.pos = u.mvp * vec4f(pos, 1.0);
  out.uv = uv;
  return out;
}
 
@fragment
fn fs(v: VSOut) -> @location(0) vec4f {
  return textureSample(tex, samp, v.uv) * u.tint;
}

const verts = new Float32Array([0, .5, -.5, -.5, .5, -.5]);
 
// 1. Create + queue upload (most common)
const vbuf = device.createBuffer({
  label: "triangle",
  size: verts.byteLength,
  usage: GPUBufferUsage.VERTEX | GPUBufferUsage.COPY_DST,
});
device.queue.writeBuffer(vbuf, 0, verts);
 
// 2. Mapped at creation (static data, no COPY_DST needed)
const ibuf = device.createBuffer({
  size: 8, // multiple of 4
  usage: GPUBufferUsage.INDEX,
  mappedAtCreation: true,
});
new Uint16Array(ibuf.getMappedRange()).set([0, 1, 2]);
ibuf.unmap();

const bgl = device.createBindGroupLayout({
  entries: [
    { binding: 0, buffer: { type: "uniform" },
      visibility: GPUShaderStage.VERTEX
        | GPUShaderStage.FRAGMENT },
    { binding: 1, texture: { sampleType: "float" },
      visibility: GPUShaderStage.FRAGMENT },
    { binding: 2, sampler: { type: "filtering" },
      visibility: GPUShaderStage.FRAGMENT },
  ],
});
const layout = device.createPipelineLayout({
  bindGroupLayouts: [bgl], // index = @group
});
const bindGroup = device.createBindGroup({
  layout: bgl,
  entries: [
    { binding: 0, resource: { buffer: ubuf } },
    { binding: 1, resource: texture.createView() },
    { binding: 2, resource: sampler },
  ],
});

const pipeline = device.createRenderPipeline({
  label: "mesh",
  layout, // or "auto"
  vertex: {
    module,
    entryPoint: "vs", // optional if the module has one
    buffers: [{
      arrayStride: 5 * 4, // bytes per vertex
      stepMode: "vertex", // "instance" for per-instance
      attributes: [
        { shaderLocation: 0, offset: 0,
          format: "float32x3" }, // pos
        { shaderLocation: 1, offset: 12,
          format: "float32x2" }, // uv
      ],
    }],
  },
  fragment: {
    module,
    entryPoint: "fs",
    targets: [{ format }],
  },
  primitive: { topology: "triangle-list", cullMode: "back" },
  depthStencil: {
    format: "depth24plus",
    depthWriteEnabled: true,
    depthCompare: "less",
  },
  multisample: { count: 1 }, // 1 or 4
});

// Straight alpha blending for a target
const blend: GPUBlendState = {
  color: {
    srcFactor: "src-alpha",
    dstFactor: "one-minus-src-alpha",
    operation: "add",
  },
  alpha: {
    srcFactor: "one",
    dstFactor: "one-minus-src-alpha",
  },
};

@group(0) @binding(0) var<storage, read> src: array<f32>;
@group(0) @binding(1) var<storage, read_write> o: array<f32>;
 
var<workgroup> tile: array<f32, 64>;
 
@compute @workgroup_size(64)
fn main(
  @builtin(global_invocation_id) gid: vec3u,
  @builtin(local_invocation_index) lid: u32,
  @builtin(workgroup_id) wid: vec3u,
) {
  let n = arrayLength(&src);
  tile[lid] = select(0.0, src[gid.x], gid.x < n);
  workgroupBarrier();
  if (lid == 0u) {
    var s = 0.0;
    for (var i = 0u; i < 64u; i++) { s += tile[i]; }
    o[wid.x] = s; // one partial sum per workgroup
  }
}

const cp = device.createComputePipeline({
  layout: "auto",
  compute: { module, constants: { /* overrides */ } },
});

function frame() {
  const encoder = device.createCommandEncoder();
  const pass = encoder.beginRenderPass({
    colorAttachments: [{
      view: context.getCurrentTexture().createView(),
      clearValue: [0.1, 0.1, 0.1, 1],
      loadOp: "clear", // or "load" to keep contents
      storeOp: "store",
    }],
  });
  pass.setPipeline(pipeline);
  pass.setBindGroup(0, bindGroup);
  pass.setVertexBuffer(0, vbuf);
  pass.draw(3); // vertexCount, instanceCount = 1
  pass.end();
  device.queue.submit([encoder.finish()]);
  requestAnimationFrame(frame);
}
requestAnimationFrame(frame);

async function loadTexture(device: GPUDevice, url: string) {
  const res = await fetch(url);
  if (!res.ok) throw new Error(`HTTP ${res.status}: ${url}`);
  const bmp = await createImageBitmap(await res.blob());
  const texture = device.createTexture({
    size: [bmp.width, bmp.height],
    format: "rgba8unorm-srgb",
    usage: GPUTextureUsage.TEXTURE_BINDING
      | GPUTextureUsage.COPY_DST
      // RENDER_ATTACHMENT is required by the copy below
      | GPUTextureUsage.RENDER_ATTACHMENT,
  });
  device.queue.copyExternalImageToTexture(
    { source: bmp }, { texture }, [bmp.width, bmp.height],
  );
  bmp.close();
  return texture;
}
 
const sampler = device.createSampler({
  magFilter: "linear",
  minFilter: "linear",
  mipmapFilter: "linear",
  addressModeU: "repeat",
  addressModeV: "repeat",
});

device.lost.then((info) => {
  console.error(`GPU lost (${info.reason})`, info.message);
  if (info.reason !== "destroyed") void restartRenderer();
});
 
device.onuncapturederror = (e) => {
  console.error("WebGPU error:", e.error.message);
};
 
// Catch errors from a specific block of calls
device.pushErrorScope("validation");
const p = device.createRenderPipeline(desc);
const err = await device.popErrorScope();
if (err) throw new Error(`Invalid pipeline: ${err.message}`);

// Shader diagnostics with line numbers
const shader = device.createShaderModule({ code });
const { messages } = await shader.getCompilationInfo();
for (const m of messages) {
  const where = `${m.lineNum}:${m.linePos}`;
  if (m.type === "error") console.error(where, m.message);
}

const code = /* wgsl */ `
@vertex fn vs(@builtin(vertex_index) i: u32)
    -> @builtin(position) vec4f {
  var p = array(vec2f(0, .5), vec2f(-.5), vec2f(.5, -.5));
  return vec4f(p[i], 0, 1); }
@fragment fn fs() -> @location(0) vec4f {
  return vec4f(1, .4, .2, 1); }`;
const { device } = await initGPU();
const canvas = document.querySelector("canvas")!;
const { context, format } = configureCanvas(canvas, device);
const module = device.createShaderModule({ code });
const pipeline = device.createRenderPipeline({
  layout: "auto",
  vertex: { module },
  fragment: { module, targets: [{ format }] },
});
const enc = device.createCommandEncoder();
const pass = enc.beginRenderPass({
  colorAttachments: [{ view: context.getCurrentTexture()
    .createView(), loadOp: "clear", storeOp: "store" }],
});
pass.setPipeline(pipeline);
pass.draw(3);
pass.end();
device.queue.submit([enc.finish()]);

export async function readBuffer(
  device: GPUDevice, src: GPUBuffer, size = src.size,
): Promise<ArrayBuffer> {
  const staging = device.createBuffer({
    size,
    usage: GPUBufferUsage.MAP_READ | GPUBufferUsage.COPY_DST,
  });
  const enc = device.createCommandEncoder();
  enc.copyBufferToBuffer(src, 0, staging, 0, size);
  device.queue.submit([enc.finish()]);
  await staging.mapAsync(GPUMapMode.READ);
  const copy = staging.getMappedRange().slice(0); // own it
  staging.unmap();
  staging.destroy();
  return copy;
}

const code = /* wgsl */ `
@group(0) @binding(0) var<storage, read_write> d: array<f32>;
@compute @workgroup_size(64) fn main(
    @builtin(global_invocation_id) id: vec3u) {
  if (id.x < arrayLength(&d)) { d[id.x] *= 2.0; } }`;
async function gpuDouble(dev: GPUDevice, xs: Float32Array) {
  const pipe = dev.createComputePipeline({ layout: "auto",
    compute: { module: dev.createShaderModule({ code }) } });
  const buf = dev.createBuffer({ size: xs.byteLength,
    mappedAtCreation: true, usage: GPUBufferUsage.STORAGE
      | GPUBufferUsage.COPY_SRC });
  new Float32Array(buf.getMappedRange()).set(xs); // upload
  buf.unmap();
  const enc = dev.createCommandEncoder();
  const pass = enc.beginComputePass();
  pass.setPipeline(pipe);
  pass.setBindGroup(0, dev.createBindGroup({
    layout: pipe.getBindGroupLayout(0),
    entries: [{ binding: 0, resource: { buffer: buf } }],
  }));
  pass.dispatchWorkgroups(Math.ceil(xs.length / 64));
  pass.end();
  dev.queue.submit([enc.finish()]);
  return new Float32Array(await readBuffer(dev, buf));
}

let depth: GPUTexture | undefined;
function frame() {
  const color = context.getCurrentTexture();
  const { width, height } = color;
  if (depth?.width !== width || depth.height !== height) {
    depth?.destroy();
    depth = device.createTexture({
      size: [width, height],
      format: "depth24plus",
      usage: GPUTextureUsage.RENDER_ATTACHMENT,
    });
  }
  const enc = device.createCommandEncoder();
  const pass = enc.beginRenderPass({
    colorAttachments: [{ view: color.createView(),
      loadOp: "clear", storeOp: "store" }],
    depthStencilAttachment: { view: depth.createView(),
      depthClearValue: 1, depthLoadOp: "clear",
      depthStoreOp: "store" },
  });
  drawScene(pass); // setPipeline, setBindGroup, draw…
  pass.end();
  device.queue.submit([enc.finish()]);
  requestAnimationFrame(frame);
}

type Backend =
  | { kind: "webgpu"; device: GPUDevice;
      context: GPUCanvasContext }
  | { kind: "webgl2"; gl: WebGL2RenderingContext };
 
async function pickBackend(
  canvas: HTMLCanvasElement,
): Promise<Backend> {
  try {
    const adapter = await navigator.gpu?.requestAdapter();
    const device = await adapter?.requestDevice();
    const context = device && canvas.getContext("webgpu");
    if (device && context) {
      context.configure({ device,
        format: navigator.gpu.getPreferredCanvasFormat() });
      return { kind: "webgpu", device, context };
    }
  } catch { /* fall through to WebGL2 */ }
  const gl = canvas.getContext("webgl2");
  if (gl) return { kind: "webgl2", gl };
  throw new Error("Neither WebGPU nor WebGL2 is available");
}