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

Don't assume all bind groups are used in create_pipeline_layout #41

Open
ScanMountGoat opened this issue May 8, 2023 · 6 comments
Open

Comments

@ScanMountGoat
Copy link
Owner

ScanMountGoat commented May 8, 2023

wgpu can detect the case where a bind group layout is not included in the pipeline layout and not used in the shader or used in the shader but not included in the pipeline layout. wgsl_to_wgpu assumes all bind groups are used, which can create problems for different pipelines in the same WGSL file. A depth only pass for shadows might not use the material uniforms buffer but still want to use the same vertex code as the color pass, for example.

This may require recreating some of the validation code used by naga or wgpu. It's also worth investigating how this works with non consecutive bind groups if index 1 is unused for bind groups 0, 1, 2.

@badicsalex
Copy link
Contributor

badicsalex commented May 10, 2023

Just a fun fact, with WGSL (I only tried compute shaders), you can have conflicting bindings as long as you don't use the conflicted ones from the same kernel:

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

@group(0) @binding(0)
var <storage, read> b:f32;

@compute @workgroup_size(32)
fn whatever1() {
    let param = a;
    ...
}

@compute @workgroup_size(32)
fn whatever2() {
    let param = b;
    ...
}

@compute @workgroup_size(32)
fn whatever2() {
    let something = a;
    // DISALLOWED:
    let param = b;
    ...
}

This is what's written in the current spec:

Two different resource variables in a shader must not have the same group and binding values, when considered as a pair.

Where a "shader" is not a script, but an entry point and everything it transitively accesses.

So if the assumption of wgsl_to_wgpu that a "shader" is the whole file, that assumption is wrong.

@ScanMountGoat
Copy link
Owner Author

We can generate different bind group initialization code for each entry point in theory. I'm not sure how easily that can be done in practice. This may also become more difficult when considering vertex+fragment shaders since each stage can access different bind groups. I would like for wgsl_to_wgpu to at least accept all valid WGSL even if the generated code has some limitations. The current bind group checking probably needs to be relaxed to conform with the spec.

@Gonkalbell
Copy link

I would like for wgsl_to_wgpu to at least accept all valid WGSL even if the generated code has some limitations. The current bind group checking probably needs to be relaxed to conform with the spec.

I definitely agree with this approach, since the spec allows you to do some pretty clever things with sharing bind groups across different pipelines, and it would be really difficult for wgsl_to_wgpu to account for them all. I've already run into several situations where wgsl_to_wgpu produced errors for shader modules that the spec allows.

I already ran into the issue that @badicsalex mentioned, in code that looked something like this:

@group(0) @binding(0) var<uniform> res_camera: Camera;
@group(1) @binding(0) var res_sampler: sampler;

// Uses res_camera
@vertex fn vs_common(...) -> ... {}

@group(1) @binding(1) var res_texture_2d: texture_2d<f32>;
// Uses res_camera, res_sampler, and res_texture_2d
@fragment fn fs_draw_texture_2d(...) -> @location(0) vec4f {...}

@group(1) @binding(1) var res_texture_cube: texture_cube<f32>;
// Uses res_camera, res_sampler, and res_texture_cube
@fragment fn fs_draw_texture_cube(...) -> @location(0) vec4f {...}

I can think of a few workarounds for this, but for me it was easier to just not use wgsl_to_wgpu.

Also, because of that todo! in quote_shader_stages, just having a compute shader in the same module as vertex/fragment shaders will panic:

@group(0) @binding(0) var storage_tex_read_write: texture_storage_2d<rgba8uint, read_write>;
@fragment fn fs_draw_texture() {}
@compute @workgroup_size(1) fn cs_update_texture() {}

I think the main problem in both of these cases is the generated bind_groups module, and by extension, the create_pipeline_layout function. It's still convenient in cases where every shader entry point in a module can use the same wgpu::PipelineLayout. But as soon as you need to mix-and-match different bind groups for different pipelines some assumptions start to break down. And it's not an easy problem to solve, since there's only so much wgsl_to_wgpu can do by statically analyzing shader files. Would it be reasonable to something to WriteOptions to opt-out of the bind_groups_module and create_pipeline_layout steps?

@ScanMountGoat
Copy link
Owner Author

Also, because of that todo! in quote_shader_stages, just having a compute shader in the same module as vertex/fragment shaders will panic

This should be fixed on the latest commit.

But as soon as you need to mix-and-match different bind groups for different pipelines some assumptions start to break down

You can always split the file into multiple WGSL files. This may require duplicating some code.

Would it be reasonable to something to WriteOptions to opt-out of the bind_groups_module and create_pipeline_layout steps

I'll look into removing the validation checks for duplicate bind groups and bindings. I don't expect all the generated code to be used by all applications.

@Gonkalbell
Copy link

This should be fixed on the latest commit.

Thanks for the quick turnaround on that! I just tried it out, but I got a compiler error from the new generated code, since the bitwise or operator isn't const:

error[E0015]: cannot call non-const operator in constants
    --> example\src\compute_shader.rs:19:25
     |
19   |             visibility: wgpu::ShaderStages::FRAGMENT | wgpu::ShaderStages::COMPUTE,
     |                         ^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^
     |
note: impl defined here, but it is not `const`

it's not pretty, but maybe you can use the .union function instead?

@ScanMountGoat
Copy link
Owner Author

That specific issue is fixed now. There are still some other issues that need to be fixed for compute entries with vertex and fragment entries to work properly.

Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment
Labels
None yet
Projects
None yet
Development

No branches or pull requests

3 participants