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

Support override-expression for workgroup_size #4450

Open
kvark opened this issue Dec 9, 2021 · 3 comments · May be fixed by #6635
Open

Support override-expression for workgroup_size #4450

kvark opened this issue Dec 9, 2021 · 3 comments · May be fixed by #6635
Labels
area: naga front-end lang: WGSL WebGPU Shading Language naga Shader Translator

Comments

@kvark
Copy link
Member

kvark commented Dec 9, 2021

Required for https://gpuharbor.ucsc.edu/

Shader '' parsing error: expected signed integer literal of 32-bit width, found 'workgroupXSize'
    ┌─ wgsl:111:34
    │
111 │ [[stage(compute), workgroup_size(workgroupXSize)]] fn main(
    │                                  ^^^^^^^^^^^^^^ expected signed integer literal of 32-bit width

, caused by: expected signed integer literal of 32-bit width, found 'workgroupXSize'
@kvark kvark added the lang: WGSL WebGPU Shading Language label Dec 9, 2021
@glalonde
Copy link
Contributor

glalonde commented Jan 23, 2022

Not sure if this is the exact same issue but it seems module scope constants are explicitly allowed for workgroup_size, but naga only accepts literals.

Would be nice to do something like this:

let WORKGROUP_SIZE: u32 = 256u;
[[stage(compute), workgroup_size(WORKGROUP_SIZE)]]
fn main([[builtin(global_invocation_id)]] global_id: vec3<u32>, [[builtin(num_workgroups)]] num_workgroups: vec3<u32>) {
    let total_particles = num_workgroups[0] * num_workgroups[1] * num_workgroups[2] * WORKGROUP_SIZE;
...

Would it work to add in a branch that uses parse_const_expression in addition to parse_generic_non_negative_int_literal? or is that more complicated than it seems

@kvark
Copy link
Member Author

kvark commented Jan 24, 2022

Yeah, this can be considered a part of the issue. Technically speaking, there should be no place in code where only literals are supported. If literals are supported, then constants as well.
This part of WGSL has been evolving recently, so we couldn't know about this when just implementing the code originally. We need to update the code now.

@cwfitzgerald cwfitzgerald transferred this issue from gfx-rs/naga Oct 25, 2023
@cwfitzgerald cwfitzgerald added the naga Shader Translator label Oct 25, 2023
@teoxoy teoxoy added this to the WebGPU Specification V1 milestone Nov 3, 2023
@jimblandy jimblandy moved this to One day in WebGPU for Firefox Dec 13, 2023
@teoxoy teoxoy removed the status in WebGPU for Firefox Dec 14, 2023
@teoxoy teoxoy changed the title Support specializing workgroup size Support override-expression for workgroup_size Jul 5, 2024
@kentslaney
Copy link

kentslaney commented Nov 28, 2024

This is the relevant line for specifying workgroup_size now (also see: [1] [2] [3])

It looks like expressions support overrides, it's just that they don't get compiled to const expressions in [1] since Module requires ast::EntryPoint to resolve into u32s for crate::EntryPoint.workgroup_size by then.

Maybe we can add a workgroup_size_overrides: Option<[Option<Handle<Expression>>; 3]> attribute to crate::EntryPoint that gets evaluated in [1] and initialize the u32s to zeros in [2] if size_expr is an ExpressionKind::Override.

I'll give it a try sometime this week.

@kentslaney kentslaney linked a pull request Nov 30, 2024 that will close this issue
7 tasks
Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment
Labels
area: naga front-end lang: WGSL WebGPU Shading Language naga Shader Translator
Projects
Status: No status
Development

Successfully merging a pull request may close this issue.

5 participants