Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Relax vector, matrix and array OOB validation #4390

Open
teoxoy opened this issue May 14, 2022 · 4 comments
Open

Relax vector, matrix and array OOB validation #4390

teoxoy opened this issue May 14, 2022 · 4 comments
Labels
area: validation Issues related to validation, diagnostics, and error handling naga Shader Translator

Comments

@teoxoy
Copy link
Member

teoxoy commented May 14, 2022

Relax vector, matrix and array OOB validation by handling OOB accesses at runtime.

The following example should work; however, validation currently errors because the index is out of bounds.

@compute @workgroup_size(1)
fn main() {
    let idx = 9;
    let v = vec2<f32>();
    let m = mat2x2<f32>();
    let a = array<f32,1>();
    let vi = v[idx];
    let mi = m[idx];
    let ai = a[idx];
}

[edit] Note that WGSL says that let-bound identifiers cannot be constant expressions, so the shader translation time bounds checking rules do not apply to this example.

Related

@teoxoy teoxoy added the area: validation Issues related to validation, diagnostics, and error handling label May 14, 2022
@cwfitzgerald cwfitzgerald added the naga Shader Translator label Oct 25, 2023
@cwfitzgerald cwfitzgerald transferred this issue from gfx-rs/naga Oct 25, 2023
@teoxoy teoxoy added this to the WebGPU Specification V1 milestone Nov 3, 2023
@jimblandy
Copy link
Member

Note that this issue only covers the excessive validation. There may be missing bounds checks, and those should be filed as separate issues.

@nical
Copy link
Contributor

nical commented Jan 3, 2024

See also: #4337

@bavalpey
Copy link

bavalpey commented Oct 3, 2024

Curious, why should this work?
Do you just mean it should pass validation? And then the bounds-check policy is used to determine what should happen with vi, mi, etc.?

I think this would be a regression. If you can detect OOB accesses, as these clearly are, at shader translation time, why not report them?

@ErichDonGubler
Copy link
Member

@bavalpey: It's for compliance with the WebGPU spec., which forces safe (but implementation-defined) behavior: https://www.w3.org/TR/WGSL/#out-of-bounds-access-sec

Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment
Labels
area: validation Issues related to validation, diagnostics, and error handling naga Shader Translator
Projects
Status: No status
Development

No branches or pull requests

6 participants