CVE-2026-9889

Critical watch CVSS 8.3 chrome OOB 📄 Reporter bug report 🔧 Commit mapped 🧬 Front-end validation exempts a record the backend still uses

Overview

Critical

Severity

8.3

CVSS

Exploited ITW

Fixed

Fix Status

ImpactOut of bounds read and write in Dawn

DescriptionOut of bounds read and write in Dawn

ComponentDawn

Sub-componentGL

Bug ClassOOB

Action tierwatch — corruption primitive but gated (uncommon) — not default-reachable
Computed by PatchGap's deterministic exploitability ruler from the fixing commit. patch-now = RCE-class & attacker-reachable in a default config, or exploited in the wild; watch = real security signal below that bar; noise = no security relevance.

Tracker511727159

Fix commit0bc97cb2f50a (dawn) +127/-0

CWECWE-125, CWE-787 (Out-of-bounds read, Out-of-bounds write)

CVSS vectorCVSS:3.1/AV:N/AC:H/PR:N/UI:R/S:C/C:H/I:H/A:H

CISA KEVNot listed

CreditedGoogle

Reported2026-05-10 (per vendor advisory)

Fix landed2026-05-21 (public git)

Disclosed2026-05-27

Patch gap6 days — fix public in source before the advisory

Background

Dawn: Chromium’s native implementation of the WebGPU API, translating WebGPU commands into calls on backend graphics APIs such as OpenGL, Vulkan, and D3D12.
`wgpu::ShaderStage::None`: A visibility flag on a bind group layout entry indicating the binding is not exposed to any shader stage, so backends may legitimately skip emitting a real GPU resource slot for it.
`GL_MAX_SHADER_STORAGE_BUFFER_BINDINGS`: The OpenGL driver limit (reported by Dawn as kGLMaxShaderStorageBufferBindingsReported) on how many SSBO binding points exist, indexing into a fixed-size internal table.
`arrayLength()` / SSBO index table: Dawn’s GL backend maintains a per-pipeline array of storage buffer binding indices used to satisfy WGSL arrayLength(), and both pipeline layout construction and bind group application must agree on those indices.

Root Cause Analysis

The OpenGL backend assigns SSBO binding indices in PipelineLayoutGL by iterating over every BindingInfo in each bind group layout and, in CommandBufferGL, applies bind groups by iterating the same bindings to bind buffers at those indices. Dawn’s front-end validation counts binding limits but deliberately excludes entries whose visibility is wgpu::ShaderStage::None, so a caller could declare far more than 96 such entries without tripping any validation. Because neither the index-assignment loop nor the bind-application loop skipped ShaderStage::None bindings, these validation-invisible entries still consumed slots in the GL binding table, pushing computed indices past kGLMaxShaderStorageBufferBindingsReported and past the bounds of the internal index array. The result was out-of-bounds reads and writes on the fixed-size table and mismatched buffer transfers.

The fix adds an early continue for bindingInfo.visibility == wgpu::ShaderStage::None in both loops, so skipped bindings no longer consume indices and the two sides of the mapping stay consistent within the driver’s reported limits.

Key insight

The single mistake was that the GL backend counted ShaderStage::None bindings when computing and applying storage-buffer indices, even though front-end validation excludes them, so index space could overflow the driver-reported maximum; the fix aligns both loops with validation by skipping those bindings entirely.

Attack Path

Declare padding bindings A page creates bind group layouts containing many entries with visibility = wgpu::ShaderStage::None, which pass validation because Dawn does not count them against binding limits.
Exceed the reported limit By spreading more than 96 such entries across several bind groups, the attacker forces the GL backend to assign SSBO indices beyond kGLMaxShaderStorageBufferBindingsReported.
Trigger index computation Creating a pipeline with this layout runs PipelineLayoutGL’s loop, which writes out-of-range indices into the fixed-size binding table.
Apply the bind groups Submitting a compute or render pass invokes CommandBufferGL’s apply loop, which reads and binds using those out-of-bounds indices, causing OOB reads and writes and corrupted buffer contents.

Impact Assessment

An attacker running WebGPU content in the GPU process on an OpenGL backend gains an out-of-bounds read and write against Dawn’s internal storage-buffer index table, corrupting adjacent memory and misrouting buffer transfers. Exploitation requires only a WebGPU-capable page on a GL-backed device and no special permissions, making it a remotely reachable memory-safety flaw in the GPU process.

Changed Functions

Function	Change	Notes
`if` src/dawn/native/opengl/CommandBufferGL.cpp	modified
`if` src/dawn/native/opengl/PipelineLayoutGL.cpp	modified
`GLArrayLengthOverflowTest` src/dawn/tests/end2end/OpArrayLengthTests.cpp	modified
`TEST_P` src/dawn/tests/end2end/OpArrayLengthTests.cpp	modified
`for` src/dawn/tests/end2end/OpArrayLengthTests.cpp	modified

Files Changed

src/dawn/native/opengl/CommandBufferGL.cpp
src/dawn/native/opengl/PipelineLayoutGL.cpp
src/dawn/tests/end2end/OpArrayLengthTests.cpp

Audit Directions

Validation-skipped entries in backends
Audit every backend loop that assigns or applies resource indices to ensure it skips exactly the entries that front-end validation excludes, especially ShaderStage::None visibility bindings.
Fixed-size backend tables vs. reported limits
Look for internal arrays sized by driver-reported maxima (like kGLMaxShaderStorageBufferBindingsReported) that are indexed by counts the front end does not bound identically.
Layout/apply loop symmetry
Check that any pair of loops which must produce matching indices (pipeline layout construction and command buffer bind application) apply identical skip and counting rules to avoid index drift and OOB access.

Patch

view commit

From 0bc97cb2f50a092b4f332c8cb2b3285de7ccc39d Mon Sep 17 00:00:00 2001
From: Stephen White <senorblanco@chromium.org>
Date: Thu, 21 May 2026 10:10:38 -0700
Subject: [PATCH] GL: fix issue with ShaderStage::None bindings.

If more than 96 bindings with visibility ShaderStage::None are used,
they will exceed kGLMaxShaderStorageBufferBindingsReported (aka
GL_MAX_SHADER_STORAGE_BUFFER_BINDINGS) in the GL backend.
Dawn validation will not catch this, since it skips such bindings.
The fix is to skip all of the bindings with visibility ShaderStage::None
when computing binding indices in pipeline layout construction and
when applying bind groups in the GL backend.

Bug: 511727159
Change-Id: Ibc873a5afab4e73c12ad25b7d7cedcf48f35ca6e
Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/309876
Reviewed-by: Corentin Wallez <cwallez@chromium.org>
Commit-Queue: Stephen White <senorblanco@chromium.org>
---

diff --git a/src/dawn/native/opengl/CommandBufferGL.cpp b/src/dawn/native/opengl/CommandBufferGL.cpp
index 72448b7..ccd52e3 100644
--- a/src/dawn/native/opengl/CommandBufferGL.cpp
+++ b/src/dawn/native/opengl/CommandBufferGL.cpp
@@ -349,6 +349,9 @@
 
         for (BindingIndex bindingIndex : Range(group->GetLayout()->GetBindingCount())) {
             const BindingInfo& bindingInfo = group->GetLayout()->GetBindingInfo(bindingIndex);
+            if (bindingInfo.visibility == wgpu::ShaderStage::None) {
+                continue;
+            }
             DAWN_TRY(MatchVariant(
                 bindingInfo.bindingLayout,
                 [&](const BufferBindingInfo& layout) -> MaybeError {
diff --git a/src/dawn/native/opengl/PipelineLayoutGL.cpp b/src/dawn/native/opengl/PipelineLayoutGL.cpp
index 98bfff0..8479add 100644
--- a/src/dawn/native/opengl/PipelineLayoutGL.cpp
+++ b/src/dawn/native/opengl/PipelineLayoutGL.cpp
@@ -48,6 +48,9 @@
 
         for (BindingIndex bindingIndex{0}; bindingIndex < bgl->GetBindingCount(); ++bindingIndex) {
             const BindingInfo& bindingInfo = bgl->GetBindingInfo(bindingIndex);
+            if (bindingInfo.visibility == wgpu::ShaderStage::None) {
+                continue;
+            }
             MatchVariant(
                 bindingInfo.bindingLayout,
                 [&](const BufferBindingInfo& layout) {
diff --git a/src/dawn/tests/end2end/OpArrayLengthTests.cpp b/src/dawn/tests/end2end/OpArrayLengthTests.cpp
index e1b4179..10b5ed4 100644
--- a/src/dawn/tests/end2end/OpArrayLengthTests.cpp
+++ b/src/dawn/tests/end2end/OpArrayLengthTests.cpp
@@ -432,5 +432,126 @@
                          VulkanBackend(), WebGPUBackend()},
                         {TieredLimits::No, TieredLimits::Yes});
 
+class GLArrayLengthOverflowTest : public DawnTest {
+  protected:
+    void GetRequiredLimits(const dawn::utils::ComboLimits& supported,
+                           dawn::utils::ComboLimits& required) override {
+        supported.UnlinkedCopyTo(&required);
+    }
+};
+
+// Test that using more than 96 ShaderStage::None bind group entries
+// (which don't count against Dawn's validation limit) don't cause GL
+// errors and failed buffer transfers.
+TEST_P(GLArrayLengthOverflowTest, VisibilityNoneOverflowsArrayLengthBuffer) {
+    DAWN_TEST_UNSUPPORTED_IF(GetSupportedLimits().maxBindGroups < 4);
+
+    constexpr uint32_t kLargeSize = 512u;
+    constexpr uint32_t kSmallSize = 256u;
+    constexpr uint32_t kPadPerGroup = 33;
+
+    wgpu::BufferDescriptor bd;
+    bd.size = kLargeSize;
+    bd.usage = wgpu::BufferUsage::Storage | wgpu::BufferUsage::CopySrc;
+    wgpu::Buffer largeBuf = device.CreateBuffer(&bd);
+
+    // Get the arrayLength() of the passed-in storage buffer, and store it into
+    // the first element of the array.
+    wgpu::ComputePipelineDescriptor primeDesc;
+    primeDesc.compute.module = utils::CreateShaderModule(device, R"(
+        @group(0) @binding(0) var<storage, read_write> a : array<u32>;
+        @compute @workgroup_size(1) fn main() {
+            a[0] = arrayLength(&a);
+        })");
+    wgpu::ComputePipeline primePipeline = device.CreateComputePipeline(&primeDesc);
+    wgpu::BindGroupLayout primeBGL = primePipeline.GetBindGroupLayout(0);
+    wgpu::BindGroup primeBG = utils::MakeBindGroup(device, primeBGL, {{0, largeBuf}});
+
+    {
+        wgpu::CommandEncoder enc = device.CreateCommandEncoder();
+        wgpu::ComputePassEncoder pass = enc.BeginComputePass();
+        pass.SetPipeline(primePipeline);
+        pass.SetBindGroup(0, primeBG);
+        pass.DispatchWorkgroups(1);
+        pass.End();
+        wgpu::CommandBuffer cb = enc.Finish();
+        queue.Submit(1, &cb);
+    }
+
+    bd.size = kSmallSize;
+    bd.usage = wgpu::BufferUsage::Storage | wgpu::BufferUsage::CopySrc;
+    wgpu::Buffer smallBuf = device.CreateBuffer(&bd);
+
+    bd.size = 4;
+    bd.usage = wgpu::BufferUsage::Storage;
+    wgpu::Buffer tinyBuffer = device.CreateBuffer(&bd);
+
+    wgpu::BindGroupLayout bgl0 = utils::MakeBindGroupLayout(
+        device, {{0, wgpu::ShaderStage::Compute, wgpu::BufferBindingType::Storage}});
+
+    std::vector<wgpu::BindGroupLayoutEntry> padEntries(kPadPerGroup);
+    for (uint32_t i = 0; i < kPadPerGroup; i++) {
+        padEntries[i].binding = i;
+        padEntries[i].visibility = wgpu::ShaderStage::None;
+        padEntries[i].buffer.type = wgpu::BufferBindingType::ReadOnlyStorage;
+    }
+    wgpu::BindGroupLayoutDescriptor padDesc;
+    padDesc.entryCount = padEntries.size();
+    padDesc.entries = padEntries.data();
+    wgpu::BindGroupLayout bglPad = device.CreateBindGroupLayout(&padDesc);
+
+    wgpu::BindGroupLayout bgls[] = {bgl0, bglPad, bglPad, bglPad};
+    wgpu::PipelineLayoutDescriptor plDesc;
+    plDesc.bindGroupLayoutCount = 4;
+    plDesc.bindGroupLayouts = bgls;
+    wgpu::PipelineLayout manyBindingsPL = device.CreatePipelineLayout(&plDesc);
+
+    wgpu::ComputePipelineDescriptor manyBindingsDesc;
+    manyBindingsDesc.layout = manyBindingsPL;
+    manyBindingsDesc.compute.module = primeDesc.compute.module;
+    wgpu::ComputePipeline manyBindingsPipeline = device.CreateComputePipeline(&manyBindingsDesc);
+
+    wgpu::BindGroup bg0 = utils::MakeBindGroup(device, bgl0, {{0, smallBuf}});
+
+    std::vector<wgpu::BindGroupEntry> padBinds(kPadPerGroup);
+    for (uint32_t i = 0; i < kPadPerGroup; i++) {
+        padBinds[i].binding = i;
+        padBinds[i].buffer = tinyBuffer;
+    }
+    wgpu::BindGroupDescriptor bgPadDesc;
+    bgPadDesc.layout = bglPad;
+    bgPadDesc.entryCount = padBinds.size();
+    bgPadDesc.entries = padBinds.data();
+    wgpu::BindGroup bgPad = device.CreateBindGroup(&bgPadDesc);
+
+    {
+        wgpu::CommandEncoder enc = device.CreateCommandEncoder();
+        wgpu::ComputePassEncoder pass = enc.BeginComputePass();
+        pass.SetPipeline(manyBindingsPipeline);
+        pass.SetBindGroup(0, bg0);
+        pass.SetBindGroup(1, bgPad);
+        pass.SetBindGroup(2, bgPad);
+        pass.SetBindGroup(3, bgPad);
+        pass.DispatchWorkgroups(1);
+        pass.End();
+        wgpu::CommandBuffer cb = enc.Finish();
+        queue.Submit(1, &cb);
+    }
+
+    // Check that the stored arrayLength is the (new) small buffer length
+    // and not the (stale) large buffer length.
+    EXPECT_BUFFER_U32_EQ(kSmallSize / 4u, smallBuf, 0);
+}
+
+DAWN_INSTANTIATE_TEST(GLArrayLengthOverflowTest,
+                      D3D11Backend(),
+                      D3D12Backend(),
+                      MetalBackend(),
+                      OpenGLBackend(),
+                      OpenGLESBackend(),
+                      OpenGLESBackend({"gl_use_array_length_from_uniform"}),
+                      VulkanBackend(),
+                      WebGPUBackend());
+
 }  // anonymous namespace
 }  // namespace dawn

Loading diff…

Regression Test / PoC

shipped with the fix

diff --git a/src/dawn/tests/end2end/OpArrayLengthTests.cpp b/src/dawn/tests/end2end/OpArrayLengthTests.cpp
index e1b4179..10b5ed4 100644
--- a/src/dawn/tests/end2end/OpArrayLengthTests.cpp
+++ b/src/dawn/tests/end2end/OpArrayLengthTests.cpp
@@ -432,5 +432,126 @@
                          VulkanBackend(), WebGPUBackend()},
                         {TieredLimits::No, TieredLimits::Yes});
 
+class GLArrayLengthOverflowTest : public DawnTest {
+  protected:
+    void GetRequiredLimits(const dawn::utils::ComboLimits& supported,
+                           dawn::utils::ComboLimits& required) override {
+        supported.UnlinkedCopyTo(&required);
+    }
+};
+
+// Test that using more than 96 ShaderStage::None bind group entries
+// (which don't count against Dawn's validation limit) don't cause GL
+// errors and failed buffer transfers.
+TEST_P(GLArrayLengthOverflowTest, VisibilityNoneOverflowsArrayLengthBuffer) {
+    DAWN_TEST_UNSUPPORTED_IF(GetSupportedLimits().maxBindGroups < 4);
+
+    constexpr uint32_t kLargeSize = 512u;
+    constexpr uint32_t kSmallSize = 256u;
+    constexpr uint32_t kPadPerGroup = 33;
+
+    wgpu::BufferDescriptor bd;
+    bd.size = kLargeSize;
+    bd.usage = wgpu::BufferUsage::Storage | wgpu::BufferUsage::CopySrc;
+    wgpu::Buffer largeBuf = device.CreateBuffer(&bd);
+
+    // Get the arrayLength() of the passed-in storage buffer, and store it into
+    // the first element of the array.
+    wgpu::ComputePipelineDescriptor primeDesc;
+    primeDesc.compute.module = utils::CreateShaderModule(device, R"(
+        @group(0) @binding(0) var<storage, read_write> a : array<u32>;
+        @compute @workgroup_size(1) fn main() {
+            a[0] = arrayLength(&a);
+        })");
+    wgpu::ComputePipeline primePipeline = device.CreateComputePipeline(&primeDesc);
+    wgpu::BindGroupLayout primeBGL = primePipeline.GetBindGroupLayout(0);
+    wgpu::BindGroup primeBG = utils::MakeBindGroup(device, primeBGL, {{0, largeBuf}});
+
+    {
+        wgpu::CommandEncoder enc = device.CreateCommandEncoder();
+        wgpu::ComputePassEncoder pass = enc.BeginComputePass();
+        pass.SetPipeline(primePipeline);
+        pass.SetBindGroup(0, primeBG);
+        pass.DispatchWorkgroups(1);
+        pass.End();
+        wgpu::CommandBuffer cb = enc.Finish();
+        queue.Submit(1, &cb);
+    }
+
+    bd.size = kSmallSize;
+    bd.usage = wgpu::BufferUsage::Storage | wgpu::BufferUsage::CopySrc;
+    wgpu::Buffer smallBuf = device.CreateBuffer(&bd);
+
+    bd.size = 4;
+    bd.usage = wgpu::BufferUsage::Storage;
+    wgpu::Buffer tinyBuffer = device.CreateBuffer(&bd);
+
+    wgpu::BindGroupLayout bgl0 = utils::MakeBindGroupLayout(
+        device, {{0, wgpu::ShaderStage::Compute, wgpu::BufferBindingType::Storage}});
+
+    std::vector<wgpu::BindGroupLayoutEntry> padEntries(kPadPerGroup);
+    for (uint32_t i = 0; i < kPadPerGroup; i++) {
+        padEntries[i].binding = i;
+        padEntries[i].visibility = wgpu::ShaderStage::None;
+        padEntries[i].buffer.type = wgpu::BufferBindingType::ReadOnlyStorage;
+    }
+    wgpu::BindGroupLayoutDescriptor padDesc;
+    padDesc.entryCount = padEntries.size();
+    padDesc.entries = padEntries.data();
+    wgpu::BindGroupLayout bglPad = device.CreateBindGroupLayout(&padDesc);
+
+    wgpu::BindGroupLayout bgls[] = {bgl0, bglPad, bglPad, bglPad};
+    wgpu::PipelineLayoutDescriptor plDesc;
+    plDesc.bindGroupLayoutCount = 4;
+    plDesc.bindGroupLayouts = bgls;
+    wgpu::PipelineLayout manyBindingsPL = device.CreatePipelineLayout(&plDesc);
+
+    wgpu::ComputePipelineDescriptor manyBindingsDesc;
+    manyBindingsDesc.layout = manyBindingsPL;
+    manyBindingsDesc.compute.module = primeDesc.compute.module;
+    wgpu::ComputePipeline manyBindingsPipeline = device.CreateComputePipeline(&manyBindingsDesc);
+
+    wgpu::BindGroup bg0 = utils::MakeBindGroup(device, bgl0, {{0, smallBuf}});
+
+    std::vector<wgpu::BindGroupEntry> padBinds(kPadPerGroup);
+    for (uint32_t i = 0; i < kPadPerGroup; i++) {
+        padBinds[i].binding = i;
+        padBinds[i].buffer = tinyBuffer;
+    }
+    wgpu::BindGroupDescriptor bgPadDesc;
+    bgPadDesc.layout = bglPad;
+    bgPadDesc.entryCount = padBinds.size();
+    bgPadDesc.entries = padBinds.data();
+    wgpu::BindGroup bgPad = device.CreateBindGroup(&bgPadDesc);
+
+    {
+        wgpu::CommandEncoder enc = device.CreateCommandEncoder();
+        wgpu::ComputePassEncoder pass = enc.BeginComputePass();
+        pass.SetPipeline(manyBindingsPipeline);
+        pass.SetBindGroup(0, bg0);
+        pass.SetBindGroup(1, bgPad);
+        pass.SetBindGroup(2, bgPad);
+        pass.SetBindGroup(3, bgPad);
+        pass.DispatchWorkgroups(1);
+        pass.End();
+        wgpu::CommandBuffer cb = enc.Finish();
+        queue.Submit(1, &cb);
+    }
+
+    // Check that the stored arrayLength is the (new) small buffer length
+    // and not the (stale) large buffer length.
+    EXPECT_BUFFER_U32_EQ(kSmallSize / 4u, smallBuf, 0);
+}
+
+DAWN_INSTANTIATE_TEST(GLArrayLengthOverflowTest,
+                      D3D11Backend(),
+                      D3D12Backend(),
+                      MetalBackend(),
+                      OpenGLBackend(),
+                      OpenGLESBackend(),
+                      OpenGLESBackend({"gl_use_array_length_from_uniform"}),
+                      VulkanBackend(),
+                      WebGPUBackend());
+
 }  // anonymous namespace
 }  // namespace dawn

Loading diff…

Original Bug Report

reported by vm...@google.com

Dawn GLES: Bypassed SSBO limit via ShaderStage::None allows OOB GPU memory access

Project Fortify, an experimental security project, has identified the following potential security issue. If you’re a feature owner CC-ed on this bug, please do your best to review these reports without the Chrome Security team. Please see https://chromium.googlesource.com/chromium/src/+/main/docs/security/ai-generated-security-bugs-faq.md for more information.

Overview: Storage buffer bindings with wgpu::ShaderStage::None visibility bypass per-stage limit validation, allowing a pipeline to exceed internal SSBO capacity. This inflates internal SSBO indices, causing a silent failure when uploading length data to an internal uniform buffer. The stale or zeroed buffer values cause Tint’s robustness transforms to underflow, resulting in an ineffective clamp and allowing arbitrary OOB GPU memory reads/writes.

Affected files:

third_party/dawn/src/dawn/native/BindingInfo.cpp
third_party/dawn/src/dawn/native/opengl/PipelineLayoutGL.cpp
third_party/dawn/src/dawn/native/opengl/CommandBufferGL.cpp
third_party/dawn/src/dawn/native/opengl/DeviceGL.cpp
third_party/dawn/src/dawn/native/opengl/PipelineLayoutGL.h
third_party/dawn/src/dawn/native/opengl/ShaderModuleGL.cpp
third_party/dawn/src/dawn/native/opengl/BufferGL.cpp
third_party/dawn/src/dawn/native/opengl/UtilsGL.h
third_party/dawn/src/tint/lang/core/ir/transform/array_length_from_uniform.cc

Estimated timestamp from git blame: 2025-12-04

Summary

A potential vulnerability in Dawn’s OpenGL/GLES backend allows for out-of-bounds (OOB) reads and writes in the GPU process.

The issue stems from a combination of factors:

A validation gap allows the creation of a BindGroupLayout with an excessive number of storage buffers by setting their visibility to wgpu::ShaderStage::None.
The OpenGL backend blindly assigns an internal ssboIndex to all storage buffers, causing the index to exceed the internal tracker’s maximum capacity (96).
Updates to the internal uniform buffer holding SSBO array lengths fail silently due to an ignored GL_INVALID_VALUE error in Release builds.
When the internal uniform buffer retains stale or out-of-bounds (zero) length values, integer underflows in Tint’s robustness and array length transforms result in massive bounds clamps, bypassing security checks.

Technical Details and Potential Exploitation Steps

Our codebase investigator agent analyzed the flow and suggests the following steps could trigger the vulnerability:

Prime the Uniform Buffer (Optional but possible): An attacker dispatches a valid compute shader using a normal layout, binding a maximum-sized storage buffer. This populates Dawn’s internal mArrayLengthBuffer (a 384-byte buffer designed to hold up to 96 SSBO lengths) with very large length values on the GPU.
Bypass Validation with ShaderStage::None: The attacker constructs a malicious BindGroupLayoutDescriptor containing over 96 dummy storage buffer bindings. By setting the visibility of these dummy bindings to wgpu::ShaderStage::None, they bypass the maxStorageBuffersPerShaderStage limit (16). In BindingInfo.cpp, IncrementBindingCounts skips incrementing the stage counters for None visibility (IterateStages(0) is empty). The total bindings easily fit within the maxBindingsPerBindGroup limit (1000).
Inflate ssboIndex: The attacker creates a PipelineLayout with this layout. The PipelineLayoutGL constructor sequentially assigns an ssboIndex to every storage buffer binding, inflating the index of the target storage buffer far past 96.
Silent Buffer Upload Failure: The attacker dispatches a malicious shader that performs OOB dynamic array indexing on the target storage buffer.
- CommandBufferGL::ApplyBindGroup calculates an update range for mArrayLengthBuffer based on the inflated ssboIndex.
- ApplyInternalArrayLengthUniforms calls glBufferSubData with an update size that exceeds the physically allocated 384 bytes of mArrayLengthBuffer.
- The OpenGL driver returns GL_INVALID_VALUE, aborting the upload. However, because the call is wrapped in DAWN_GL_TRY, the error is silently ignored in Release builds (where DAWN_ENABLE_ASSERTS is disabled). The uniform buffer on the GPU is not updated.
Robustness Bypass via Underflow: The malicious shader executes. Because the uniform buffer wasn’t updated, a read from the out-of-bounds index returns 0 (due to standard GL robustness rules).
- In array_length_from_uniform.cc, the array size is calculated as total_buffer_size - member->Offset(). If total_buffer_size is 0, this unsigned subtraction underflows to a massive value.
- Even if the offset is 0, robustness.cc calculates the clamp limit as arrayLength() - 1. 0 - 1 underflows to 0xFFFFFFFF.
- The clamp operation becomes min(index, 0xFFFFFFFF), which effectively disables the clamp, allowing the attacker-controlled index to be used directly for OOB GPU memory reads and writes.

Impact

This vulnerability provides a potential path to arbitrary shader-driven out-of-bounds read and write capabilities in the GPU virtual address space. On Android platforms where the GPU process is frequently unsandboxed, this could lead to a sandbox escape and Remote Code Execution (RCE) from unprivileged Web content.

Suggested Fixes

Validation: In BindingInfo.cpp, enforce a hard, cross-stage limit on the total number of storage buffer bindings allowed in a PipelineLayout, regardless of shader stage visibility. Do not rely solely on per-stage accumulated counts.
Error Handling: Ensure that failures during internal buffer synchronization (e.g., glBufferSubData in ApplyInternalArrayLengthUniforms) cause the dispatch to fail safely, rather than silently ignoring the error via DAWN_GL_TRY in non-assert builds.
Tint Robustness: Guard against integer underflows in array_length_from_uniform.cc and robustness.cc. Ensure that if a buffer size is smaller than the array offset (or exactly 0), the resulting clamped array length does not wrap around to a massive maximum value.

Evaluated with Chrome root at commit: eca8648a4e1cdfdda68c495a6003059fed641955

Results so far have been promising, but there can be wrong deductions. If this proves to be a false positive, please close as WAI; data from false positives will be used to improve accuracy over time. And please feel free to reach out to me directly if you have concerns or feedback on the project.

View on issue tracker

References

Vendor advisory Fix commit Bug tracker NVD

On This Page