WebGPU 的新变化 (Chrome 123)

François Beaufort
François Beaufort
GitHub

WGSL 中的 DP4a 内置函数支持

DP4a（4 个元素的点积与累加）指的是一组 GPU 指令，常用于深度学习推理量化。它可以高效地执行 8 位整数点积，以加快此类 int8 量化模型的计算。与 f32 版本相比，它可以节省（高达 75%）的内存和网络带宽，并提升任何机器学习模型在推断方面的性能。因此，它现在在许多流行的 AI 框架中得到大量使用。

现在，当 navigator.gpu.wgslLanguageFeatures 中存在 "packed_4x8_integer_dot_product" WGSL 语言扩展时，您可以通过 dot4U8Packed 和 dot4I8Packed 内置函数，使用将 8 位整数的 4 分量向量打包为 WGSL 着色器代码中的点积指令的 32 位整数标量。您还可以通过 pack4xI8、pack4xU8、pack4xI8Clamp、pack4xU8Clamp、unpack4xI8 和 unpack4xU8 WGSL 内置函数将打包和解压缩指令与 8 位整数的打包 4 分量向量结合使用。

建议使用 WGSL 着色器代码顶部的 requires packed_4x8_integer_dot_product; 来指示不可移植的不可移植性。详见下例和 issue tint:1497。

if (!navigator.gpu.wgslLanguageFeatures.has("packed_4x8_integer_dot_product")) {
  throw new Error(`DP4a built-in functions are not available`);
}

const adapter = await navigator.gpu.requestAdapter();
const device = await adapter.requestDevice();

const shaderModule = device.createShaderModule({ code: `
  requires packed_4x8_integer_dot_product;

  fn main() {
    const result: u32 = dot4U8Packed(0x01020304u, 0x02040405u); // 42
  }`,
});

特别感谢 Intel 上海网络图形团队推动此规范和实施的顺利完成！

WGSL 中的指针参数不受限制

"unrestricted_pointer_parameters" WGSL 语言扩展放宽了对可以向 WGSL 函数传递哪些指针的限制：

  • storage、uniform 和 workgroup 地址空间的参数指针，指向用户声明的函数。

  • 将指向结构成员和数组元素的指针传递给用户声明的函数。

如需了解详情，请参阅指针即函数参数 | WGSL 导览。

可以使用 navigator.gpu.wgslLanguageFeatures 检测此功能。建议始终在 WGSL 着色器代码顶部使用 requires unrestricted_pointer_parameters; 来指示不可移植的不可移植性。请参阅以下示例、WGSL 规范变更和 issue tint:2053。

if (!navigator.gpu.wgslLanguageFeatures.has("unrestricted_pointer_parameters")) {
  throw new Error(`Unrestricted pointer parameters are not available`);
}

const adapter = await navigator.gpu.requestAdapter();
const device = await adapter.requestDevice();

const shaderModule = device.createShaderModule({ code: `
  requires unrestricted_pointer_parameters;

  @group(0) @binding(0) var<storage, read_write> S : i32;

  fn func(pointer : ptr<storage, i32, read_write>) {
    *pointer = 42;
  }

  @compute @workgroup_size(1)
  fn main() {
    func(&S);
  }`
});

用于在 WGSL 中解引用组合的语法糖

当 navigator.gpu.wgslLanguageFeatures 中存在 "pointer_composite_access" WGSL 语言扩展时，您的 WGSL 着色器代码现在支持使用同一点 (.) 语法访问复杂数据类型的组件，无论您是直接使用数据还是使用指向数据的指针。具体方法如下：

  • 如果 foo 是指针，则使用 foo.bar 编写 (*foo).bar 更方便。通常需要使用星号 (*) 才能将指针转换为可以解引用的“引用”，但现在指针和引用更加相似并且几乎可以互换。

  • 如果 foo 不是指针：点 (.) 运算符与您通常用于直接访问成员的方式完全相同。

同样，如果 pa 是用于存储数组起始地址的指针，则可以使用 pa[i] 直接访问存储该数组的 'i 元素的内存位置。

建议使用 WGSL 着色器代码顶部的 requires pointer_composite_access; 来指示不可移植的不可移植性。详见下例和 issue tint:2113。

if (!navigator.gpu.wgslLanguageFeatures.has("pointer_composite_access")) {
  throw new Error(`Pointer composite access is not available`);
}

const adapter = await navigator.gpu.requestAdapter();
const device = await adapter.requestDevice();

const shaderModule = device.createShaderModule({ code: `
  requires pointer_composite_access;

  fn main() {
    var a = vec3f();
    let p : ptr<function, vec3f> = &a;
    let r1 = (*p).x; // always valid.
    let r2 = p.x; // requires pointer composite access.
  }`
});

模板和深度切面的单独只读状态

以前，渲染通道中的只读深度模板附件需要同时要求两个方面（深度和模板）均为只读状态。已解除此限制。现在，您可以只读方式使用深度切面，例如用于接触阴影追踪，而通过写入模板缓冲区来识别像素以进行进一步处理。请参阅问题 dawn:2146。

黎明动态

现在，在发生错误时，系统会立即调用使用 wgpuDeviceSetUncapturedErrorCallback() 设置的未捕获错误回调函数。这是开发者始终期望和期望的调试结果。请参阅更改 dawn:173620。

实现了 webgpu.h API 中的 wgpuSurfaceGetPreferredFormat() 方法。请参阅问题 dawn:1362。

这仅涵盖了部分重要的亮点。查看详尽的提交内容列表。

WebGPU 的新变化

WebGPU 的新变化系列中涵盖的所有内容的列表。

Chrome 125

Chrome 124

Chrome 123

Chrome 122

Chrome 121

Chrome 120

Chrome 119

Chrome 118

Chrome 117

Chrome 116

Chrome 115

Chrome 114

Chrome 113