From 84e9ccee6c6aed347fa0c2a207c967e41f1f8b98 Mon Sep 17 00:00:00 2001
From: Jiawei Shao <jiawei.shao@intel.com>
Date: Thu, 14 May 2026 14:17:48 +0800
Subject: [PATCH 1/2] Add shader validation tests on `subgroup-size-control` -
 Part I

This patch adds the first part of the shader validation tests on
the extension `subgroup-size-control`:
- Enabling `subgroup_size_control` must also enable `subgroups`.
- The WGSL attribute `@subgroup_size` must be used when
  `subgroup_size_control` is enabled.
- `@subgroup_size` can only be used in the compute stage

issue: #4640
---
 .../extension/subgroup_size_control.spec.ts   | 129 ++++++++++++++++++
 1 file changed, 129 insertions(+)
 create mode 100644 src/webgpu/shader/validation/extension/subgroup_size_control.spec.ts

diff --git a/src/webgpu/shader/validation/extension/subgroup_size_control.spec.ts b/src/webgpu/shader/validation/extension/subgroup_size_control.spec.ts
new file mode 100644
index 000000000000..656d3aac215d
--- /dev/null
+++ b/src/webgpu/shader/validation/extension/subgroup_size_control.spec.ts
@@ -0,0 +1,129 @@
+export const description = `
+Validation tests for the subgroup_size_control extension
+`;
+
+import { makeTestGroup } from '../../../../common/framework/test_group.js';
+import { UniqueFeaturesAndLimitsShaderValidationTest } from '../shader_validation_test.js';
+
+export const g = makeTestGroup(UniqueFeaturesAndLimitsShaderValidationTest);
+
+/**
+ * Returns a subgroup size value that is valid for use in the @subgroup_size
+ * attribute on the current adapter.
+ *
+ * On Intel gen-12lp, subgroupMinSize may be 8 in fragment stages, which is below the allowed range
+ * for `[WaveSize]` on D3D12 (can only be 16). subgroupMaxSize (16) is always within the explicit
+ * range, so it is returned for that architecture.
+ * On all other adapters, subgroupMinSize is returned as the conservative choice as on many D3D12
+ * drivers only `waveLaneCountMin` is reliable, while `waveLaneCountMax` is not.
+ *
+ * @param adapterInfo The GPUAdapterInfo of the current device's adapter.
+ * @returns A power-of-two subgroup size valid for @subgroup_size on this adapter.
+ */
+export function getValidSubgroupSizeForSubgroupSizeAttribute(adapterInfo: GPUAdapterInfo): number {
+  interface SubgroupAdapterInfo extends GPUAdapterInfo {
+    subgroupMinSize: number;
+    subgroupMaxSize: number;
+  }
+  const { vendor, architecture, subgroupMinSize, subgroupMaxSize } =
+    adapterInfo as SubgroupAdapterInfo;
+  return vendor === 'intel' && architecture === 'gen-12lp' ? subgroupMaxSize : subgroupMinSize;
+}
+
+g.test('enable_subgroup_size_control_requires_subgroups')
+  .desc(
+    `Checks that enabling the WGSL extension subgroup_size_control without also enabling the
+     subgroups extension is a compilation error. The test is skipped if the adapter does not
+     support the subgroup-size-control WebGPU extension.`
+  )
+  .params(u => u.combine('enableSubgroups', [false, true] as const))
+  .beforeAllSubcases(t => {
+    t.selectDeviceOrSkipTestCase({
+      requiredFeatures: ['subgroup-size-control' as GPUFeatureName],
+    });
+  })
+  .fn(t => {
+    const { enableSubgroups } = t.params;
+
+    t.expectCompileResult(
+      enableSubgroups,
+      `
+        ${enableSubgroups ? 'enable subgroups;' : ''}
+        enable subgroup_size_control;
+        @compute @workgroup_size(1)
+        fn main() {}
+      `
+    );
+  });
+
+g.test('use_subgroup_size_attribute_requires_subgroup_size_control_extension_enabled')
+  .desc(
+    `Checks that the @subgroup_size attribute is only allowed with the WGSL extension
+     subgroup_size_control enabled in the shader and the WebGPU extension subgroup-size-control
+     supported on the device. The test is skipped if the adapter does not support the
+     subgroup-size-control WebGPU extension.`
+  )
+  .params(u => u.combine('enableExtension', [false, true] as const))
+  .beforeAllSubcases(t => {
+    t.selectDeviceOrSkipTestCase({
+      requiredFeatures: ['subgroup-size-control' as GPUFeatureName],
+    });
+  })
+  .fn(t => {
+    const { enableExtension } = t.params;
+
+    const subgroupSize = getValidSubgroupSizeForSubgroupSizeAttribute(t.device.adapterInfo);
+    t.expectCompileResult(
+      enableExtension,
+      `
+        ${enableExtension ? 'enable subgroups; enable subgroup_size_control;' : ''}
+        @compute @workgroup_size(${subgroupSize}) @subgroup_size(${subgroupSize})
+        fn main() {}
+      `
+    );
+  });
+
+const kStageShaders = {
+  compute: (subgroupSize: number) => `
+    enable subgroups;
+    enable subgroup_size_control;
+    @compute @workgroup_size(${subgroupSize}) @subgroup_size(${subgroupSize})
+    fn main() {}
+  `,
+  vertex: (subgroupSize: number) => `
+    enable subgroups;
+    enable subgroup_size_control;
+    @vertex @subgroup_size(${subgroupSize})
+    fn main() -> @builtin(position) vec4f {
+      return vec4f(0);
+    }
+  `,
+  fragment: (subgroupSize: number) => `
+    enable subgroups;
+    enable subgroup_size_control;
+    @fragment @subgroup_size(${subgroupSize})
+    fn main() -> @location(0) vec4f {
+      return vec4f(0);
+    }
+  `,
+} as const;
+
+g.test('subgroup_size_attribute_only_valid_in_compute_stage')
+  .desc(
+    `Checks that the @subgroup_size attribute is only valid on a compute shader entry point.
+     Applying it to a vertex or fragment entry point must be a compilation error.
+     The test is skipped if the adapter does not support the subgroup-size-control WebGPU
+     extension.`
+  )
+  .params(u => u.combine('stage', ['compute', 'vertex', 'fragment'] as const))
+  .beforeAllSubcases(t => {
+    t.selectDeviceOrSkipTestCase({
+      requiredFeatures: ['subgroup-size-control' as GPUFeatureName],
+    });
+  })
+  .fn(t => {
+    const { stage } = t.params;
+    const subgroupSize = getValidSubgroupSizeForSubgroupSizeAttribute(t.device.adapterInfo);
+
+    t.expectCompileResult(stage === 'compute', kStageShaders[stage](subgroupSize));
+  });

From 1756791175aa5d9735850e865ee60553ff4d8012 Mon Sep 17 00:00:00 2001
From: Jiawei Shao <jiawei.shao@intel.com>
Date: Thu, 14 May 2026 14:28:47 +0800
Subject: [PATCH 2/2] Fix comments

---
 .../validation/extension/subgroup_size_control.spec.ts | 10 +++-------
 1 file changed, 3 insertions(+), 7 deletions(-)

diff --git a/src/webgpu/shader/validation/extension/subgroup_size_control.spec.ts b/src/webgpu/shader/validation/extension/subgroup_size_control.spec.ts
index 656d3aac215d..34597b531fad 100644
--- a/src/webgpu/shader/validation/extension/subgroup_size_control.spec.ts
+++ b/src/webgpu/shader/validation/extension/subgroup_size_control.spec.ts
@@ -33,8 +33,7 @@ export function getValidSubgroupSizeForSubgroupSizeAttribute(adapterInfo: GPUAda
 g.test('enable_subgroup_size_control_requires_subgroups')
   .desc(
     `Checks that enabling the WGSL extension subgroup_size_control without also enabling the
-     subgroups extension is a compilation error. The test is skipped if the adapter does not
-     support the subgroup-size-control WebGPU extension.`
+     subgroups extension is a compilation error.`
   )
   .params(u => u.combine('enableSubgroups', [false, true] as const))
   .beforeAllSubcases(t => {
@@ -60,8 +59,7 @@ g.test('use_subgroup_size_attribute_requires_subgroup_size_control_extension_ena
   .desc(
     `Checks that the @subgroup_size attribute is only allowed with the WGSL extension
      subgroup_size_control enabled in the shader and the WebGPU extension subgroup-size-control
-     supported on the device. The test is skipped if the adapter does not support the
-     subgroup-size-control WebGPU extension.`
+     supported on the device.`
   )
   .params(u => u.combine('enableExtension', [false, true] as const))
   .beforeAllSubcases(t => {
@@ -111,9 +109,7 @@ const kStageShaders = {
 g.test('subgroup_size_attribute_only_valid_in_compute_stage')
   .desc(
     `Checks that the @subgroup_size attribute is only valid on a compute shader entry point.
-     Applying it to a vertex or fragment entry point must be a compilation error.
-     The test is skipped if the adapter does not support the subgroup-size-control WebGPU
-     extension.`
+     Applying it to a vertex or fragment entry point must be a compilation error.`
   )
   .params(u => u.combine('stage', ['compute', 'vertex', 'fragment'] as const))
   .beforeAllSubcases(t => {