diff --git a/src/amd/compiler/aco_instruction_selection.cpp b/src/amd/compiler/aco_instruction_selection.cpp index 2faee4ce9d5..a1ecc410ddd 100644 --- a/src/amd/compiler/aco_instruction_selection.cpp +++ b/src/amd/compiler/aco_instruction_selection.cpp @@ -8181,7 +8181,7 @@ visit_intrinsic(isel_context* ctx, nir_intrinsic_instr* instr) case nir_intrinsic_scoped_barrier: emit_scoped_barrier(ctx, instr); break; case nir_intrinsic_load_num_workgroups: { Temp dst = get_ssa_temp(ctx, &instr->dest.ssa); - if (ctx->options->load_grid_size_from_user_sgpr) { + if (ctx->args->load_grid_size_from_user_sgpr) { bld.copy(Definition(dst), get_arg(ctx, ctx->args->ac.num_work_groups)); } else { Temp addr = get_arg(ctx, ctx->args->ac.num_work_groups); diff --git a/src/amd/vulkan/radv_nir_to_llvm.c b/src/amd/vulkan/radv_nir_to_llvm.c index 32c22a4159f..2bb4a9d77c1 100644 --- a/src/amd/vulkan/radv_nir_to_llvm.c +++ b/src/amd/vulkan/radv_nir_to_llvm.c @@ -2311,7 +2311,7 @@ ac_translate_nir_to_llvm(struct ac_llvm_compiler *ac_llvm, ctx.abi.adjust_frag_coord_z = options->adjust_frag_coord_z; ctx.abi.robust_buffer_access = options->robust_buffer_access; ctx.abi.disable_aniso_single_level = options->disable_aniso_single_level; - ctx.abi.load_grid_size_from_user_sgpr = options->load_grid_size_from_user_sgpr; + ctx.abi.load_grid_size_from_user_sgpr = args->load_grid_size_from_user_sgpr; bool is_ngg = is_pre_gs_stage(shaders[0]->info.stage) && info->is_ngg; if (shader_count >= 2 || is_ngg) diff --git a/src/amd/vulkan/radv_pipeline.c b/src/amd/vulkan/radv_pipeline.c index 958395b1233..2bc5e66de11 100644 --- a/src/amd/vulkan/radv_pipeline.c +++ b/src/amd/vulkan/radv_pipeline.c @@ -37,6 +37,7 @@ #include "radv_meta.h" #include "radv_private.h" #include "radv_shader.h" +#include "radv_shader_args.h" #include "vk_util.h" #include "util/debug.h" @@ -3297,6 +3298,56 @@ radv_fill_shader_info(struct radv_pipeline *pipeline, } } +static void +radv_declare_pipeline_args(struct radv_device *device, struct radv_shader_args *args, + nir_shader **nir, struct radv_shader_info *infos, + const struct radv_pipeline_key *pipeline_key) +{ + enum chip_class chip_class = device->physical_device->rad_info.chip_class; + unsigned active_stages = 0; + + for (int i = 0; i < MESA_VULKAN_SHADER_STAGES; i++) { + if (nir[i]) + active_stages |= (1 << i); + } + + for (int i = 0; i < MESA_VULKAN_SHADER_STAGES; ++i) { + args[i].is_gs_copy_shader = false; + args[i].explicit_scratch_args = !radv_use_llvm_for_stage(device, i); + args[i].remap_spi_ps_input = !radv_use_llvm_for_stage(device, i); + args[i].load_grid_size_from_user_sgpr = device->load_grid_size_from_user_sgpr; + } + + if (chip_class >= GFX9 && nir[MESA_SHADER_TESS_CTRL]) { + radv_declare_shader_args(chip_class, pipeline_key, &infos[MESA_SHADER_TESS_CTRL], + MESA_SHADER_TESS_CTRL, true, MESA_SHADER_VERTEX, + &args[MESA_SHADER_TESS_CTRL]); + infos[MESA_SHADER_TESS_CTRL].user_sgprs_locs = args[MESA_SHADER_TESS_CTRL].user_sgprs_locs; + + args[MESA_SHADER_VERTEX] = args[MESA_SHADER_TESS_CTRL]; + active_stages &= ~(1 << MESA_SHADER_VERTEX); + active_stages &= ~(1 << MESA_SHADER_TESS_CTRL); + } + + if (chip_class >= GFX9 && nir[MESA_SHADER_GEOMETRY]) { + gl_shader_stage pre_stage = + nir[MESA_SHADER_TESS_EVAL] ? MESA_SHADER_TESS_EVAL : MESA_SHADER_VERTEX; + radv_declare_shader_args(chip_class, pipeline_key, &infos[MESA_SHADER_GEOMETRY], + MESA_SHADER_GEOMETRY, true, pre_stage, &args[MESA_SHADER_GEOMETRY]); + infos[MESA_SHADER_GEOMETRY].user_sgprs_locs = args[MESA_SHADER_GEOMETRY].user_sgprs_locs; + + args[pre_stage] = args[MESA_SHADER_GEOMETRY]; + active_stages &= ~(1 << pre_stage); + active_stages &= ~(1 << MESA_SHADER_GEOMETRY); + } + + u_foreach_bit(i, active_stages) { + radv_declare_shader_args(chip_class, pipeline_key, &infos[i], i, false, MESA_SHADER_VERTEX, + &args[i]); + infos[i].user_sgprs_locs = args[i].user_sgprs_locs; + } +} + static void merge_tess_info(struct shader_info *tes_info, struct shader_info *tcs_info) { @@ -3995,6 +4046,9 @@ radv_create_shaders(struct radv_pipeline *pipeline, struct radv_pipeline_layout radv_determine_ngg_settings(pipeline, pipeline_key, infos, nir); + struct radv_shader_args args[MESA_VULKAN_SHADER_STAGES] = {{{{{0}}}}}; + radv_declare_pipeline_args(device, args, nir, infos, pipeline_key); + for (int i = 0; i < MESA_VULKAN_SHADER_STAGES; ++i) { if (nir[i]) { radv_start_feedback(stage_feedbacks[i]); @@ -4116,9 +4170,16 @@ radv_create_shaders(struct radv_pipeline *pipeline, struct radv_pipeline_layout info.workgroup_size = 64; /* HW VS: separate waves, no workgroups */ info.ballot_bit_size = 64; + struct radv_shader_args gs_copy_args = {0}; + gs_copy_args.is_gs_copy_shader = true; + gs_copy_args.explicit_scratch_args = !radv_use_llvm_for_stage(device, MESA_SHADER_VERTEX); + radv_declare_shader_args(device->physical_device->rad_info.chip_class, pipeline_key, &info, + MESA_SHADER_VERTEX, false, MESA_SHADER_VERTEX, &gs_copy_args); + info.user_sgprs_locs = gs_copy_args.user_sgprs_locs; + pipeline->gs_copy_shader = radv_create_gs_copy_shader( - device, nir[MESA_SHADER_GEOMETRY], &info, &gs_copy_binary, keep_executable_info, - keep_statistic_info, pipeline_key->has_multiview_view_index, + device, nir[MESA_SHADER_GEOMETRY], &info, &gs_copy_args, &gs_copy_binary, + keep_executable_info, keep_statistic_info, pipeline_key->has_multiview_view_index, pipeline_key->optimisations_disabled); } @@ -4128,8 +4189,8 @@ radv_create_shaders(struct radv_pipeline *pipeline, struct radv_pipeline_layout pipeline->shaders[MESA_SHADER_FRAGMENT] = radv_shader_compile( device, modules[MESA_SHADER_FRAGMENT], &nir[MESA_SHADER_FRAGMENT], 1, pipeline_layout, - pipeline_key, infos + MESA_SHADER_FRAGMENT, keep_executable_info, - keep_statistic_info, &binaries[MESA_SHADER_FRAGMENT]); + pipeline_key, infos + MESA_SHADER_FRAGMENT, &args[MESA_SHADER_FRAGMENT], + keep_executable_info, keep_statistic_info, &binaries[MESA_SHADER_FRAGMENT]); radv_stop_feedback(stage_feedbacks[MESA_SHADER_FRAGMENT], false); } @@ -4143,8 +4204,8 @@ radv_create_shaders(struct radv_pipeline *pipeline, struct radv_pipeline_layout pipeline->shaders[MESA_SHADER_TESS_CTRL] = radv_shader_compile( device, modules[MESA_SHADER_TESS_CTRL], combined_nir, 2, pipeline_layout, pipeline_key, - &infos[MESA_SHADER_TESS_CTRL], keep_executable_info, keep_statistic_info, - &binaries[MESA_SHADER_TESS_CTRL]); + &infos[MESA_SHADER_TESS_CTRL], &args[MESA_SHADER_TESS_CTRL], keep_executable_info, + keep_statistic_info, &binaries[MESA_SHADER_TESS_CTRL]); radv_stop_feedback(stage_feedbacks[MESA_SHADER_TESS_CTRL], false); } @@ -4161,7 +4222,7 @@ radv_create_shaders(struct radv_pipeline *pipeline, struct radv_pipeline_layout pipeline->shaders[MESA_SHADER_GEOMETRY] = radv_shader_compile( device, modules[MESA_SHADER_GEOMETRY], combined_nir, 2, pipeline_layout, pipeline_key, - &infos[MESA_SHADER_GEOMETRY], keep_executable_info, + &infos[MESA_SHADER_GEOMETRY], &args[MESA_SHADER_GEOMETRY], keep_executable_info, keep_statistic_info, &binaries[MESA_SHADER_GEOMETRY]); radv_stop_feedback(stage_feedbacks[MESA_SHADER_GEOMETRY], false); @@ -4174,7 +4235,7 @@ radv_create_shaders(struct radv_pipeline *pipeline, struct radv_pipeline_layout radv_start_feedback(stage_feedbacks[i]); pipeline->shaders[i] = radv_shader_compile( - device, modules[i], &nir[i], 1, pipeline_layout, pipeline_key, infos + i, + device, modules[i], &nir[i], 1, pipeline_layout, pipeline_key, infos + i, &args[i], keep_executable_info, keep_statistic_info, &binaries[i]); radv_stop_feedback(stage_feedbacks[i], false); diff --git a/src/amd/vulkan/radv_shader.c b/src/amd/vulkan/radv_shader.c index 146b1d4821d..eab1485cf87 100644 --- a/src/amd/vulkan/radv_shader.c +++ b/src/amd/vulkan/radv_shader.c @@ -1907,9 +1907,10 @@ radv_dump_nir_shaders(struct nir_shader *const *shaders, int shader_count) static struct radv_shader * shader_compile(struct radv_device *device, struct vk_shader_module *module, struct nir_shader *const *shaders, int shader_count, gl_shader_stage stage, - struct radv_shader_info *info, struct radv_nir_compiler_options *options, - bool gs_copy_shader, bool trap_handler_shader, bool keep_shader_info, - bool keep_statistic_info, struct radv_shader_binary **binary_out) + struct radv_shader_info *info, const struct radv_shader_args *args, + struct radv_nir_compiler_options *options, bool gs_copy_shader, + bool trap_handler_shader, bool keep_shader_info, bool keep_statistic_info, + struct radv_shader_binary **binary_out) { enum radeon_family chip_family = device->physical_device->rad_info.family; struct radv_shader_binary *binary = NULL; @@ -1934,38 +1935,26 @@ shader_compile(struct radv_device *device, struct vk_shader_module *module, module && !is_meta_shader(module->nir) && options->key.ps.enable_mrt_output_nan_fixup; options->adjust_frag_coord_z = options->key.adjust_frag_coord_z; options->disable_aniso_single_level = options->key.disable_aniso_single_level; - options->load_grid_size_from_user_sgpr = device->load_grid_size_from_user_sgpr; options->has_image_load_dcc_bug = device->physical_device->rad_info.has_image_load_dcc_bug; options->debug.func = radv_compiler_debug; options->debug.private_data = &debug_data; - struct radv_shader_args args = {0}; - args.is_gs_copy_shader = gs_copy_shader; - args.is_trap_handler_shader = trap_handler_shader; - - radv_declare_shader_args(options, info, - gs_copy_shader ? MESA_SHADER_VERTEX : shaders[shader_count - 1]->info.stage, - shader_count >= 2, - shader_count >= 2 ? shaders[shader_count - 2]->info.stage : MESA_SHADER_VERTEX, &args); - - info->user_sgprs_locs = args.user_sgprs_locs; - #ifdef LLVM_AVAILABLE if (radv_use_llvm_for_stage(device, stage) || options->dump_shader || options->record_ir) ac_init_llvm_once(); if (radv_use_llvm_for_stage(device, stage)) { - llvm_compile_shader(options, info, shader_count, shaders, &binary, &args); + llvm_compile_shader(options, info, shader_count, shaders, &binary, args); #else if (false) { #endif } else { - aco_compile_shader(options, info, shader_count, shaders, &args, &binary); + aco_compile_shader(options, info, shader_count, shaders, args, &binary); } binary->info = *info; - struct radv_shader *shader = radv_shader_create(device, binary, keep_shader_info, false, &args); + struct radv_shader *shader = radv_shader_create(device, binary, keep_shader_info, false, args); if (!shader) { free(binary); return NULL; @@ -2004,10 +1993,9 @@ shader_compile(struct radv_device *device, struct vk_shader_module *module, struct radv_shader * radv_shader_compile(struct radv_device *device, struct vk_shader_module *module, struct nir_shader *const *shaders, int shader_count, - struct radv_pipeline_layout *layout, - const struct radv_pipeline_key *key, - struct radv_shader_info *info, bool keep_shader_info, - bool keep_statistic_info, + struct radv_pipeline_layout *layout, const struct radv_pipeline_key *key, + struct radv_shader_info *info, const struct radv_shader_args *args, + bool keep_shader_info, bool keep_statistic_info, struct radv_shader_binary **binary_out) { gl_shader_stage stage = shaders[shader_count - 1]->info.stage; @@ -2017,30 +2005,26 @@ radv_shader_compile(struct radv_device *device, struct vk_shader_module *module, if (key) options.key = *key; - options.explicit_scratch_args = !radv_use_llvm_for_stage(device, stage); - options.remap_spi_ps_input = !radv_use_llvm_for_stage(device, stage); options.robust_buffer_access = device->robust_buffer_access; options.wgp_mode = radv_should_use_wgp_mode(device, stage, info); - return shader_compile(device, module, shaders, shader_count, stage, info, &options, false, false, - keep_shader_info, keep_statistic_info, binary_out); + return shader_compile(device, module, shaders, shader_count, stage, info, args, &options, false, + false, keep_shader_info, keep_statistic_info, binary_out); } struct radv_shader * radv_create_gs_copy_shader(struct radv_device *device, struct nir_shader *shader, - struct radv_shader_info *info, struct radv_shader_binary **binary_out, - bool keep_shader_info, bool keep_statistic_info, bool multiview, - bool disable_optimizations) + struct radv_shader_info *info, const struct radv_shader_args *args, + struct radv_shader_binary **binary_out, bool keep_shader_info, + bool keep_statistic_info, bool multiview, bool disable_optimizations) { struct radv_nir_compiler_options options = {0}; gl_shader_stage stage = MESA_SHADER_VERTEX; - options.explicit_scratch_args = !radv_use_llvm_for_stage(device, stage); - options.remap_spi_ps_input = !radv_use_llvm_for_stage(device, stage); options.key.has_multiview_view_index = multiview; options.key.optimisations_disabled = disable_optimizations; - return shader_compile(device, NULL, &shader, 1, stage, info, &options, true, false, + return shader_compile(device, NULL, &shader, 1, stage, info, args, &options, true, false, keep_shader_info, keep_statistic_info, binary_out); } @@ -2051,6 +2035,7 @@ radv_create_trap_handler_shader(struct radv_device *device) struct radv_shader *shader = NULL; struct radv_shader_binary *binary = NULL; struct radv_shader_info info = {0}; + struct radv_pipeline_key key = {0}; struct radv_trap_handler_shader *trap; trap = malloc(sizeof(struct radv_trap_handler_shader)); @@ -2059,12 +2044,17 @@ radv_create_trap_handler_shader(struct radv_device *device) nir_builder b = radv_meta_init_shader(MESA_SHADER_COMPUTE, "meta_trap_handler"); - options.explicit_scratch_args = true; options.wgp_mode = radv_should_use_wgp_mode(device, MESA_SHADER_COMPUTE, &info); info.wave_size = 64; - shader = shader_compile(device, NULL, &b.shader, 1, MESA_SHADER_COMPUTE, &info, &options, false, - true, true, false, &binary); + struct radv_shader_args args; + args.explicit_scratch_args = true; + args.is_trap_handler_shader = true; + radv_declare_shader_args(device->physical_device->rad_info.chip_class, &key, &info, + MESA_SHADER_COMPUTE, false, MESA_SHADER_VERTEX, &args); + + shader = shader_compile(device, NULL, &b.shader, 1, MESA_SHADER_COMPUTE, &info, &args, &options, + false, true, true, false, &binary); trap->alloc = radv_alloc_shader_memory(device, shader->code_size, NULL); @@ -2131,8 +2121,8 @@ upload_vs_prolog(struct radv_device *device, struct radv_prolog_binary *bin, uns struct radv_shader_prolog * radv_create_vs_prolog(struct radv_device *device, const struct radv_vs_prolog_key *key) { + struct radv_shader_args args = {0}; struct radv_nir_compiler_options options = {0}; - options.explicit_scratch_args = true; options.family = device->physical_device->rad_info.family; options.chip_class = device->physical_device->rad_info.chip_class; options.info = &device->physical_device->rad_info; @@ -2151,9 +2141,11 @@ radv_create_vs_prolog(struct radv_device *device, const struct radv_vs_prolog_ke info.vs.as_ls = key->as_ls; info.is_ngg = key->is_ngg; - struct radv_shader_args args = {0}; - radv_declare_shader_args(&options, &info, key->next_stage, key->next_stage != MESA_SHADER_VERTEX, - MESA_SHADER_VERTEX, &args); + struct radv_pipeline_key pipeline_key = {0}; + + args.explicit_scratch_args = true; + radv_declare_shader_args(options.chip_class, &pipeline_key, &info, key->next_stage, + key->next_stage != MESA_SHADER_VERTEX, MESA_SHADER_VERTEX, &args); info.user_sgprs_locs = args.user_sgprs_locs; diff --git a/src/amd/vulkan/radv_shader.h b/src/amd/vulkan/radv_shader.h index 2282caff8ab..afb487980af 100644 --- a/src/amd/vulkan/radv_shader.h +++ b/src/amd/vulkan/radv_shader.h @@ -47,6 +47,7 @@ struct radv_pipeline; struct radv_pipeline_cache; struct radv_pipeline_key; struct radv_vs_input_state; +struct radv_shader_args; enum radv_vs_input_alpha_adjust { ALPHA_ADJUST_NONE = 0, @@ -113,7 +114,6 @@ enum radv_compiler_debug_level { struct radv_nir_compiler_options { struct radv_pipeline_layout *layout; struct radv_pipeline_key key; - bool explicit_scratch_args; bool robust_buffer_access; bool adjust_frag_coord_z; bool dump_shader; @@ -125,9 +125,7 @@ struct radv_nir_compiler_options { bool has_image_load_dcc_bug; bool enable_mrt_output_nan_fixup; bool wgp_mode; - bool remap_spi_ps_input; bool disable_aniso_single_level; - bool load_grid_size_from_user_sgpr; enum radeon_family family; enum chip_class chip_class; const struct radeon_info *info; @@ -539,8 +537,8 @@ struct radv_shader *radv_shader_create(struct radv_device *device, struct radv_shader *radv_shader_compile( struct radv_device *device, struct vk_shader_module *module, struct nir_shader *const *shaders, int shader_count, struct radv_pipeline_layout *layout, const struct radv_pipeline_key *key, - struct radv_shader_info *info, bool keep_shader_info, bool keep_statistic_info, - struct radv_shader_binary **binary_out); + struct radv_shader_info *info, const struct radv_shader_args *args, bool keep_shader_info, + bool keep_statistic_info, struct radv_shader_binary **binary_out); bool radv_shader_binary_upload(struct radv_device *device, const struct radv_shader_binary *binary, struct radv_shader *shader, void *dest_ptr); @@ -551,8 +549,9 @@ void radv_free_shader_memory(struct radv_device *device, union radv_shader_arena struct radv_shader * radv_create_gs_copy_shader(struct radv_device *device, struct nir_shader *nir, - struct radv_shader_info *info, struct radv_shader_binary **binary_out, - bool multiview, bool keep_shader_info, bool keep_statistic_info, + struct radv_shader_info *info, const struct radv_shader_args *args, + struct radv_shader_binary **binary_out, bool multiview, + bool keep_shader_info, bool keep_statistic_info, bool disable_optimizations); struct radv_trap_handler_shader *radv_create_trap_handler_shader(struct radv_device *device); diff --git a/src/amd/vulkan/radv_shader_args.c b/src/amd/vulkan/radv_shader_args.c index a6ff6085777..bd40dfe8d3a 100644 --- a/src/amd/vulkan/radv_shader_args.c +++ b/src/amd/vulkan/radv_shader_args.c @@ -75,19 +75,17 @@ struct user_sgpr_info { }; static bool -needs_view_index_sgpr(const struct radv_nir_compiler_options *options, - const struct radv_shader_info *info, gl_shader_stage stage) +needs_view_index_sgpr(const struct radv_pipeline_key *key, const struct radv_shader_info *info, + gl_shader_stage stage) { switch (stage) { case MESA_SHADER_VERTEX: if (info->uses_view_index || - (!info->vs.as_es && !info->vs.as_ls && - options->key.has_multiview_view_index)) + (!info->vs.as_es && !info->vs.as_ls && key->has_multiview_view_index)) return true; break; case MESA_SHADER_TESS_EVAL: - if (info->uses_view_index || - (!info->tes.as_es && options->key.has_multiview_view_index)) + if (info->uses_view_index || (!info->tes.as_es && key->has_multiview_view_index)) return true; break; case MESA_SHADER_TESS_CTRL: @@ -95,12 +93,11 @@ needs_view_index_sgpr(const struct radv_nir_compiler_options *options, return true; break; case MESA_SHADER_GEOMETRY: - if (info->uses_view_index || - (info->is_ngg && options->key.has_multiview_view_index)) + if (info->uses_view_index || (info->is_ngg && key->has_multiview_view_index)) return true; break; case MESA_SHADER_MESH: - if (info->uses_view_index || options->key.has_multiview_view_index) + if (info->uses_view_index || key->has_multiview_view_index) return true; break; default: @@ -190,10 +187,10 @@ allocate_inline_push_consts(const struct radv_shader_info *info, } static void -allocate_user_sgprs(const struct radv_nir_compiler_options *options, - const struct radv_shader_info *info, gl_shader_stage stage, - bool has_previous_stage, gl_shader_stage previous_stage, bool needs_view_index, - bool has_api_gs, bool is_gs_copy_shader, struct user_sgpr_info *user_sgpr_info) +allocate_user_sgprs(enum chip_class chip_class, const struct radv_shader_info *info, + struct radv_shader_args *args, gl_shader_stage stage, bool has_previous_stage, + gl_shader_stage previous_stage, bool needs_view_index, bool has_api_gs, + struct user_sgpr_info *user_sgpr_info) { uint8_t user_sgpr_count = 0; @@ -211,14 +208,14 @@ allocate_user_sgprs(const struct radv_nir_compiler_options *options, if (info->cs.uses_sbt) user_sgpr_count += 1; if (info->cs.uses_grid_size) - user_sgpr_count += options->load_grid_size_from_user_sgpr ? 3 : 2; + user_sgpr_count += args->load_grid_size_from_user_sgpr ? 3 : 2; if (info->cs.uses_ray_launch_size) user_sgpr_count += 3; break; case MESA_SHADER_FRAGMENT: break; case MESA_SHADER_VERTEX: - if (!is_gs_copy_shader) + if (!args->is_gs_copy_shader) user_sgpr_count += count_vs_user_sgprs(info); break; case MESA_SHADER_TESS_CTRL: @@ -257,8 +254,7 @@ allocate_user_sgprs(const struct radv_nir_compiler_options *options, if (info->so.num_outputs) user_sgpr_count++; - uint32_t available_sgprs = - options->chip_class >= GFX9 && stage != MESA_SHADER_COMPUTE ? 32 : 16; + uint32_t available_sgprs = chip_class >= GFX9 && stage != MESA_SHADER_COMPUTE ? 32 : 16; uint32_t remaining_sgprs = available_sgprs - user_sgpr_count; uint32_t num_desc_set = util_bitcount(info->desc_set_used_mask); @@ -329,14 +325,14 @@ declare_vs_specific_input_sgprs(const struct radv_shader_info *info, struct radv } static void -declare_vs_input_vgprs(const struct radv_nir_compiler_options *options, - const struct radv_shader_info *info, struct radv_shader_args *args) +declare_vs_input_vgprs(enum chip_class chip_class, const struct radv_shader_info *info, + struct radv_shader_args *args) { ac_add_arg(&args->ac, AC_ARG_VGPR, 1, AC_ARG_INT, &args->ac.vertex_id); if (!args->is_gs_copy_shader) { if (info->vs.as_ls) { ac_add_arg(&args->ac, AC_ARG_VGPR, 1, AC_ARG_INT, &args->ac.vs_rel_patch_id); - if (options->chip_class >= GFX10) { + if (chip_class >= GFX10) { ac_add_arg(&args->ac, AC_ARG_VGPR, 1, AC_ARG_INT, NULL); /* user vgpr */ ac_add_arg(&args->ac, AC_ARG_VGPR, 1, AC_ARG_INT, &args->ac.instance_id); } else { @@ -344,7 +340,7 @@ declare_vs_input_vgprs(const struct radv_nir_compiler_options *options, ac_add_arg(&args->ac, AC_ARG_VGPR, 1, AC_ARG_INT, NULL); /* unused */ } } else { - if (options->chip_class >= GFX10) { + if (chip_class >= GFX10) { if (info->is_ngg) { ac_add_arg(&args->ac, AC_ARG_VGPR, 1, AC_ARG_INT, NULL); /* user vgpr */ ac_add_arg(&args->ac, AC_ARG_VGPR, 1, AC_ARG_INT, NULL); /* user vgpr */ @@ -428,8 +424,7 @@ declare_ms_input_vgprs(struct radv_shader_args *args) } static void -declare_ps_input_vgprs(const struct radv_shader_info *info, struct radv_shader_args *args, - bool remap_spi_ps_input) +declare_ps_input_vgprs(const struct radv_shader_info *info, struct radv_shader_args *args) { unsigned spi_ps_input = info->ps.spi_ps_input; @@ -450,7 +445,7 @@ declare_ps_input_vgprs(const struct radv_shader_info *info, struct radv_shader_a ac_add_arg(&args->ac, AC_ARG_VGPR, 1, AC_ARG_INT, &args->ac.sample_coverage); ac_add_arg(&args->ac, AC_ARG_VGPR, 1, AC_ARG_INT, NULL); /* fixed pt */ - if (remap_spi_ps_input) { + if (args->remap_spi_ps_input) { /* LLVM optimizes away unused FS inputs and computes spi_ps_input_addr itself and then * communicates the results back via the ELF binary. Mirror what LLVM does by re-mapping the * VGPR arguments here. @@ -553,16 +548,16 @@ set_ms_input_locs(struct radv_shader_args *args, uint8_t *user_sgpr_idx) } void -radv_declare_shader_args(const struct radv_nir_compiler_options *options, +radv_declare_shader_args(enum chip_class chip_class, const struct radv_pipeline_key *key, const struct radv_shader_info *info, gl_shader_stage stage, bool has_previous_stage, gl_shader_stage previous_stage, struct radv_shader_args *args) { struct user_sgpr_info user_sgpr_info; - bool needs_view_index = needs_view_index_sgpr(options, info, stage); + bool needs_view_index = needs_view_index_sgpr(key, info, stage); bool has_api_gs = stage == MESA_SHADER_GEOMETRY; - if (options->chip_class >= GFX10 && info->is_ngg && stage != MESA_SHADER_GEOMETRY) { + if (chip_class >= GFX10 && info->is_ngg && stage != MESA_SHADER_GEOMETRY) { /* Handle all NGG shaders as GS to simplify the code here. */ previous_stage = stage; stage = MESA_SHADER_GEOMETRY; @@ -574,10 +569,10 @@ radv_declare_shader_args(const struct radv_nir_compiler_options *options, for (int i = 0; i < AC_UD_MAX_UD; i++) args->user_sgprs_locs.shader_data[i].sgpr_idx = -1; - allocate_user_sgprs(options, info, stage, has_previous_stage, previous_stage, needs_view_index, - has_api_gs, args->is_gs_copy_shader, &user_sgpr_info); + allocate_user_sgprs(chip_class, info, args, stage, has_previous_stage, previous_stage, + needs_view_index, has_api_gs, &user_sgpr_info); - if (options->explicit_scratch_args) { + if (args->explicit_scratch_args) { ac_add_arg(&args->ac, AC_ARG_SGPR, 2, AC_ARG_CONST_DESC_PTR, &args->ring_offsets); } @@ -594,7 +589,7 @@ radv_declare_shader_args(const struct radv_nir_compiler_options *options, } if (info->cs.uses_grid_size) { - if (options->load_grid_size_from_user_sgpr) + if (args->load_grid_size_from_user_sgpr) ac_add_arg(&args->ac, AC_ARG_SGPR, 3, AC_ARG_INT, &args->ac.num_work_groups); else ac_add_arg(&args->ac, AC_ARG_SGPR, 2, AC_ARG_CONST_PTR, &args->ac.num_work_groups); @@ -614,7 +609,7 @@ radv_declare_shader_args(const struct radv_nir_compiler_options *options, ac_add_arg(&args->ac, AC_ARG_SGPR, 1, AC_ARG_INT, &args->ac.tg_size); } - if (options->explicit_scratch_args) { + if (args->explicit_scratch_args) { ac_add_arg(&args->ac, AC_ARG_SGPR, 1, AC_ARG_INT, &args->ac.scratch_offset); } @@ -644,11 +639,11 @@ radv_declare_shader_args(const struct radv_nir_compiler_options *options, declare_streamout_sgprs(info, args, stage); } - if (options->explicit_scratch_args) { + if (args->explicit_scratch_args) { ac_add_arg(&args->ac, AC_ARG_SGPR, 1, AC_ARG_INT, &args->ac.scratch_offset); } - declare_vs_input_vgprs(options, info, args); + declare_vs_input_vgprs(chip_class, info, args); break; case MESA_SHADER_TESS_CTRL: if (has_previous_stage) { @@ -672,7 +667,7 @@ radv_declare_shader_args(const struct radv_nir_compiler_options *options, ac_add_arg(&args->ac, AC_ARG_VGPR, 1, AC_ARG_INT, &args->ac.tcs_patch_id); ac_add_arg(&args->ac, AC_ARG_VGPR, 1, AC_ARG_INT, &args->ac.tcs_rel_ids); - declare_vs_input_vgprs(options, info, args); + declare_vs_input_vgprs(chip_class, info, args); } else { declare_global_input_sgprs(info, &user_sgpr_info, args); @@ -682,7 +677,7 @@ radv_declare_shader_args(const struct radv_nir_compiler_options *options, ac_add_arg(&args->ac, AC_ARG_SGPR, 1, AC_ARG_INT, &args->ac.tess_offchip_offset); ac_add_arg(&args->ac, AC_ARG_SGPR, 1, AC_ARG_INT, &args->ac.tcs_factor_offset); - if (options->explicit_scratch_args) { + if (args->explicit_scratch_args) { ac_add_arg(&args->ac, AC_ARG_SGPR, 1, AC_ARG_INT, &args->ac.scratch_offset); } ac_add_arg(&args->ac, AC_ARG_VGPR, 1, AC_ARG_INT, &args->ac.tcs_patch_id); @@ -706,7 +701,7 @@ radv_declare_shader_args(const struct radv_nir_compiler_options *options, declare_streamout_sgprs(info, args, stage); ac_add_arg(&args->ac, AC_ARG_SGPR, 1, AC_ARG_INT, &args->ac.tess_offchip_offset); } - if (options->explicit_scratch_args) { + if (args->explicit_scratch_args) { ac_add_arg(&args->ac, AC_ARG_SGPR, 1, AC_ARG_INT, &args->ac.scratch_offset); } declare_tes_input_vgprs(args); @@ -754,7 +749,7 @@ radv_declare_shader_args(const struct radv_nir_compiler_options *options, ac_add_arg(&args->ac, AC_ARG_VGPR, 1, AC_ARG_INT, &args->ac.gs_vtx_offset[2]); if (previous_stage == MESA_SHADER_VERTEX) { - declare_vs_input_vgprs(options, info, args); + declare_vs_input_vgprs(chip_class, info, args); } else if (previous_stage == MESA_SHADER_TESS_EVAL) { declare_tes_input_vgprs(args); } else if (previous_stage == MESA_SHADER_MESH) { @@ -773,7 +768,7 @@ radv_declare_shader_args(const struct radv_nir_compiler_options *options, ac_add_arg(&args->ac, AC_ARG_SGPR, 1, AC_ARG_INT, &args->ac.gs2vs_offset); ac_add_arg(&args->ac, AC_ARG_SGPR, 1, AC_ARG_INT, &args->ac.gs_wave_id); - if (options->explicit_scratch_args) { + if (args->explicit_scratch_args) { ac_add_arg(&args->ac, AC_ARG_SGPR, 1, AC_ARG_INT, &args->ac.scratch_offset); } ac_add_arg(&args->ac, AC_ARG_VGPR, 1, AC_ARG_INT, &args->ac.gs_vtx_offset[0]); @@ -790,11 +785,11 @@ radv_declare_shader_args(const struct radv_nir_compiler_options *options, declare_global_input_sgprs(info, &user_sgpr_info, args); ac_add_arg(&args->ac, AC_ARG_SGPR, 1, AC_ARG_INT, &args->ac.prim_mask); - if (options->explicit_scratch_args) { + if (args->explicit_scratch_args) { ac_add_arg(&args->ac, AC_ARG_SGPR, 1, AC_ARG_INT, &args->ac.scratch_offset); } - declare_ps_input_vgprs(info, args, options->remap_spi_ps_input); + declare_ps_input_vgprs(info, args); break; default: unreachable("Shader stage not implemented"); @@ -823,7 +818,7 @@ radv_declare_shader_args(const struct radv_nir_compiler_options *options, } if (args->ac.num_work_groups.used) { set_loc_shader(args, AC_UD_CS_GRID_SIZE, &user_sgpr_idx, - options->load_grid_size_from_user_sgpr ? 3 : 2); + args->load_grid_size_from_user_sgpr ? 3 : 2); } if (args->ac.ray_launch_size.used) { set_loc_shader(args, AC_UD_CS_RAY_LAUNCH_SIZE, &user_sgpr_idx, 3); diff --git a/src/amd/vulkan/radv_shader_args.h b/src/amd/vulkan/radv_shader_args.h index e3b2911adac..78ab60414b2 100644 --- a/src/amd/vulkan/radv_shader_args.h +++ b/src/amd/vulkan/radv_shader_args.h @@ -50,6 +50,9 @@ struct radv_shader_args { struct radv_userdata_locations user_sgprs_locs; unsigned num_user_sgprs; + bool explicit_scratch_args; + bool remap_spi_ps_input; + bool load_grid_size_from_user_sgpr; bool is_gs_copy_shader; bool is_trap_handler_shader; }; @@ -60,10 +63,10 @@ radv_shader_args_from_ac(struct ac_shader_args *args) return container_of(args, struct radv_shader_args, ac); } -struct radv_nir_compiler_options; +struct radv_pipeline_key; struct radv_shader_info; -void radv_declare_shader_args(const struct radv_nir_compiler_options *options, +void radv_declare_shader_args(enum chip_class chip_class, const struct radv_pipeline_key *key, const struct radv_shader_info *info, gl_shader_stage stage, bool has_previous_stage, gl_shader_stage previous_stage, struct radv_shader_args *args);