Added required_subgroup_size to PipelineShaderStageCreateInfo

[charles-r-earp]

Compute and Graphics pipelines can be created with a required subgroup size.

https://registry.khronos.org/vulkan/specs/1.3-extensions/man/html/VK_EXT_subgroup_size_control.html

To do

• Validation / validation errors.
• Tests

Limitations

For compute, can't validate max_compute_workgroup_subgroups without knowing the workgroup size, which is not currently reflected.

This interacts with additional VK_EXT_subgroup_size_control flags for requiring full subgroups and allowing varying subgroups, but since those aren't currently implemented they were neglected. If implemented, they add additional constraints to validate.

[Rua]

#2234 will make some changes to how PipelineShaderStageCreateInfo is validated, so it may be better to wait until that is merged.

[Rua]

All the variables holding data to be passed to Vulkan end in _vk. Could you include that as well please?

[Rua]

#2234 has been merged now.

[Rua]

@charles-r-earp are you still interested in getting this merged?

[charles-r-earp]

@charles-r-earp are you still interested in getting this merged?

Yes

[Rua]

A few points:

• You forgot to include the format! macro when creating most of the error messages. That means that for example {subgroup_size} will appear literally in the message, not the value of the variable.
• Please put backticks around the variable and property names in the problem field (but not the context field).
• In a few places you wrote an error message instead of a VUID.
• VUID-VkPipelineShaderStageCreateInfo-pNext-02756 still needs to be validated.
• Can you fix the review points I mentioned earlier?

[charles-r-earp]

• VUID-VkPipelineShaderStageCreateInfo-pNext-02756 still needs to be validated.

This is validated:

if let Some(workgroup_size) = workgroup_size {
    let max_compute_workgroup_subgroups = properties
        .max_compute_workgroup_subgroups
        .unwrap_or_default();
    if max_compute_workgroup_subgroups * subgroup_size < workgroup_size {
        return Err(Box::new(ValidationError {
            context: "required_subgroup_size".into(),
            problem: format!("`subgroup_size` {subgroup_size} creates more than {max_compute_workgroup_subgroups} subgroups").into(),
            vuids: &["VUID-VkPipelineShaderStageCreateInfo-pNext-02756"],
            ..Default::default()
        }));
    }
}

However, workgroup_size is not available because it's not reflected from the shader, so it's currently None.

[Rua]

Can you add the necessary reflection then?

[Rua]

You can use contains_enum(stage) here and then you don't need to do the conversion above.

[Rua]

It looks good so far, but I noticed that the required_subgroup_size test is causing an overflow when I run it:

---- pipeline::compute::tests::required_subgroup_size stdout ----
thread 'pipeline::compute::tests::required_subgroup_size' panicked at 'attempt to multiply with overflow', vulkano/src/pipeline/mod.rs:582:12

It looks like max_compute_workgroup_subgroups * subgroup_size can be too big to fit into an u32. Perhaps the expression can be rearranged so that an overflow can't happen.

[charles-r-earp]

It looks like max_compute_workgroup_subgroups * subgroup_size can be too big to fit into an u32. Perhaps the expression can be rearranged so that an overflow can't happen.

While subgroup_size is up to 128, max_compute_subgroups can be u32::MAX so I decided to use checked_mul for this. But I also realized that this was a problem for local_size too because while the product of [x, y, z] must be a u32, each dimension can be u32::MAX as well, so just casting to u64 seems simpler.

[Rua]

But I also realized that this was a problem for local_size too because while the product of [x, y, z] must be a u32, each dimension can be u32::MAX as well, so just casting to u64 seems simpler.

u32::MAX * u32::MAX * u32::MAX is still bigger than u64::MAX though!

[charles-r-earp]

u32::MAX * u32::MAX * u32::MAX is still bigger than u64::MAX though!

Nice catch!

[Rua]

Looks good now, thank you!