WebGPU の新機能(Chrome 128)

François Beaufort
François Beaufort

サブグループのテスト

サブグループ機能を使用すると、SIMD レベルの並列処理が可能になり、グループ内のスレッドによる通信や、16 個の数値の合計の計算など、集団的な数学演算を実行できます。これにより、非常に効率的な形式のクロススレッド データ共有が可能になります。

「安全でない WebGPU のサポート」の背後にあるローカルテストでは、サブグループの提案の最小限の実装を利用できます。chrome://flags/#enable-unsafe-webgpuにあります

オリジン トライアルに登録すると、実際のユーザーでサイト上のサブグループを試すこともできます。オリジン トライアルを使用するためのサイトの準備方法については、オリジン トライアルのスタートガイドをご覧ください。オリジン トライアルは Chrome 128 ~ 131 まで実施されます(2025 年 2 月 19 日に終了します)。テストの意図をご覧ください。

GPUAdapter"subgroups" 機能を使用できる場合は、この機能を使用して GPUDevice をリクエストし、WGSL でサブグループのサポートを利用し、minSubgroupSizemaxSubgroupSize の制限を確認します。

また、WGSL のコードで enable subgroups; を使用して、この拡張機能を明示的に有効にする必要もあります。有効にすると、次の追加機能を利用できるようになります。

  • subgroup_invocation_id: サブグループ内のスレッドのインデックスの組み込み値。
  • subgroup_size: サブグループ サイズ アクセスの組み込み値。
  • subgroupBallot(value): アクティブな呼び出しで value が true の場合は subgroup_invocation_id に対応するビットが 1、それ以外の場合は 0 であるビット フィールドのセットを返します。
  • subgroupBroadcast(value, id): id に一致する subgroup_invocation_id の呼び出しから、サブグループ内のすべての呼び出しに value をブロードキャストします。注: id はコンパイル時の定数でなければなりません。

今後、subgroupAddsubgroupAllsubgroupElectsubgroupShuffle などの組み込み関数が追加される予定です。問題 354738715 をご覧ください。

サブグループ オペレーションで f16 を許可するには、"subgroups""subgroups-f16""shader-f16" の各機能を使用して GPUDevice をリクエストしてから、WGSL コードで enable f16, subgroups, subgroups_f16; を使用して有効にします。

次のコード スニペットは、サブグループを調整してその可能性を発見するための土台となります。

const adapter = await navigator.gpu.requestAdapter();
if (!adapter.features.has("subgroups")) {
  throw new Error("Subgroups support is not available");
}
// Explicitly request subgroups support.
const device = await adapter.requestDevice({
  requiredFeatures: ["subgroups"],
});

const shaderModule = device.createShaderModule({ code: `
  enable subgroups;

  var<workgroup> wgmem : u32;

  @group(0) @binding(0)
  var<storage, read> inputs : array<u32>;

  @group(0) @binding(1)
  var<storage, read_write> output : array<u32>;

  @compute @workgroup_size(64)
  fn main(@builtin(subgroup_size) subgroupSize : u32,
          @builtin(subgroup_invocation_id) id : u32,
          @builtin(local_invocation_index) lid : u32) {
    // One thread per workgroup writes the value to workgroup memory.
    if (lid == 0) {
      wgmem = inputs[lid];
    }
    workgroupBarrier();
    var v = 0u;

    // One thread per subgroup reads the value from workgroup memory
    // and shares that value with every other thread in the subgroup
    // to reduce local memory bandwidth.
    if (id == 0) {
      v = wgmem;
    }
    v = subgroupBroadcast(v, 0);
    output[lid] = v;
  }`,
});

// Send the appropriate commands to the GPU...

ラインとポイントの深度バイアス設定のサポート終了

WebGPU の仕様の変更により、レンダリング パイプラインのトポロジが線型または点型の場合、depthBiasdepthBiasSlopeScaledepthBiasClamp をゼロ以外の値に設定すると、検証エラーになります。デベロッパーがコードを更新する十分な時間を確保するために、この予定の検証に関する警告が DevTools コンソールに表示されます。また、そのような状況では値が強制的に 0 に設定されます。問題 352567424 をご覧ください。

未キャプチャ エラーの DevTools の警告を PreventDefault で非表示にする

DevTools コンソールで、uncapturederror のイベント リスナーが登録されていて、イベント リスナー コールバック内でイベントの preventDefault() メソッドが呼び出された場合、uncapturederror イベントに関する警告が表示されなくなりました。この動作は、JavaScript でのイベント処理と一致します。次の例と問題 40263619 をご覧ください。

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

device.addEventListener("uncapturederror", (event) => {
  // Prevents browser warning to show up in the DevTools Console.
  event.preventDefault();

  // TODO: Handle event.error
});

WGSL はまずサンプリングを補間し、

WGSL の interpolate 属性を使用すると、ユーザー定義の IO データの補間を管理できます。新たに補間サンプリング パラメータ first(デフォルト)と either が追加され、制御が強化されました。first ではプリミティブの最初の頂点の値が使用され、either では最初または最後の頂点のいずれかが許可されます。問題 340278447 をご覧ください。

夜明けの最新情報

非同期処理を処理する Dawn の WGPUFuture の実装が完了しました。主なコンセプトには、日和見イベント処理のための wgpuInstanceProcessEvents やコールバックの場所を定義する WGPUCallbackMode などがあります。WGPUFuture は、存続期間が無限の 1 回限りのイベントを示し、wgpuInstanceWaitAny は Future の完了またはタイムアウトを待機します。問題 42240932 をご覧ください。

CompositeAlphaMode::Auto 値は Surface::GetCapabilities() によって報告されなくなりました。これも引き続き有効であり、Surface::GetCapabilities().alphaMode[0] と同等です。問題 292 をご覧ください。

OpenGL バックエンドで、Present() 呼び出しごとに y フリップ ブリットを使用する Surface がサポートされるようになりました。問題 344814083 をご覧ください。

Adapter::GetProperties() メソッドのサポートは終了し、Adapter::GetInfo() に置き換えられました。

外部コントリビューターである Jaswant が、すべての CMake ファイルを書き換えたため、更新が簡単になり、事前ビルドが可能になりました。CMake プロジェクトで Dawn を使用する方法については、クイックスタート ガイドをご覧ください。

ここでは、主なハイライトの一部のみを取り上げています。コミットの一覧をご確認ください。

WebGPU の新機能

WebGPU の新機能」シリーズに記載されている全内容のリスト。

Chrome 128

Chrome 127

Chrome 126

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