Module: Mesa Branch: main Commit: d2754b12cd7385bf108a6f19746bf09ac13b38a5 URL: http://cgit.freedesktop.org/mesa/mesa/commit/?id=d2754b12cd7385bf108a6f19746bf09ac13b38a5
Author: Constantine Shablya <[email protected]> Date: Fri Jul 15 17:01:34 2022 +0300 radv: use nir_opt_uniform_access Reviewed-by: Jason Ekstrand <[email protected]> Reviewed-by: Bas Nieuwenhuizen <[email protected]> Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/17558> --- src/amd/vulkan/radv_pipeline.c | 28 ++++++++++++++++++++-------- 1 file changed, 20 insertions(+), 8 deletions(-) diff --git a/src/amd/vulkan/radv_pipeline.c b/src/amd/vulkan/radv_pipeline.c index 3847f983eb0..8ba9ba79944 100644 --- a/src/amd/vulkan/radv_pipeline.c +++ b/src/amd/vulkan/radv_pipeline.c @@ -4785,14 +4785,26 @@ radv_create_shaders(struct radv_pipeline *pipeline, struct radv_pipeline_layout /* Wave and workgroup size should already be filled. */ assert(stages[i].info.wave_size && stages[i].info.workgroup_size); - if (!radv_use_llvm_for_stage(device, i)) { - nir_lower_non_uniform_access_options options = { - .types = nir_lower_non_uniform_ubo_access | nir_lower_non_uniform_ssbo_access | - nir_lower_non_uniform_texture_access | nir_lower_non_uniform_image_access, - .callback = &non_uniform_access_callback, - .callback_data = NULL, - }; - NIR_PASS(_, stages[i].nir, nir_lower_non_uniform_access, &options); + enum nir_lower_non_uniform_access_type lower_non_uniform_access_types = + nir_lower_non_uniform_ubo_access | nir_lower_non_uniform_ssbo_access | + nir_lower_non_uniform_texture_access | nir_lower_non_uniform_image_access; + + /* In practice, most shaders do not have non-uniform-qualified + * accesses (see + * https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/17558#note_1475069) + * thus a cheaper and likely to fail check is run first. + */ + if (nir_has_non_uniform_access(stages[i].nir, lower_non_uniform_access_types)) { + NIR_PASS(_, stages[i].nir, nir_opt_non_uniform_access); + + if (!radv_use_llvm_for_stage(device, i)) { + nir_lower_non_uniform_access_options options = { + .types = lower_non_uniform_access_types, + .callback = &non_uniform_access_callback, + .callback_data = NULL, + }; + NIR_PASS(_, stages[i].nir, nir_lower_non_uniform_access, &options); + } } NIR_PASS(_, stages[i].nir, nir_lower_memory_model);
