Lines Matching full:nir
29 #include "nir/nir.h"
30 #include "nir/nir_builder.h"
31 #include "nir/nir_xfb_info.h"
141 is_meta_shader(nir_shader *nir) in is_meta_shader() argument
143 return nir && nir->info.internal; in is_meta_shader()
147 radv_can_dump_shader(struct radv_device *device, nir_shader *nir, bool meta_shader) in radv_can_dump_shader() argument
152 if ((is_meta_shader(nir) || meta_shader) && in radv_can_dump_shader()
160 radv_can_dump_shader_stats(struct radv_device *device, nir_shader *nir) in radv_can_dump_shader_stats() argument
163 return device->instance->debug_flags & RADV_DEBUG_DUMP_SHADER_STATS && !is_meta_shader(nir); in radv_can_dump_shader_stats()
231 radv_optimize_nir_algebraic(nir_shader *nir, bool opt_offsets) in radv_optimize_nir_algebraic() argument
236 NIR_PASS(_, nir, nir_copy_prop); in radv_optimize_nir_algebraic()
237 NIR_PASS(_, nir, nir_opt_dce); in radv_optimize_nir_algebraic()
238 NIR_PASS(_, nir, nir_opt_constant_folding); in radv_optimize_nir_algebraic()
239 NIR_PASS(_, nir, nir_opt_cse); in radv_optimize_nir_algebraic()
240 NIR_PASS(more_algebraic, nir, nir_opt_algebraic); in radv_optimize_nir_algebraic()
249 NIR_PASS(_, nir, nir_opt_offsets, &offset_options); in radv_optimize_nir_algebraic()
261 NIR_PASS(more_late_algebraic, nir, nir_opt_algebraic_late); in radv_optimize_nir_algebraic()
262 NIR_PASS(_, nir, nir_opt_constant_folding); in radv_optimize_nir_algebraic()
263 NIR_PASS(_, nir, nir_copy_prop); in radv_optimize_nir_algebraic()
264 NIR_PASS(_, nir, nir_opt_dce); in radv_optimize_nir_algebraic()
265 NIR_PASS(_, nir, nir_opt_cse); in radv_optimize_nir_algebraic()
322 lower_intrinsics(nir_shader *nir, const struct radv_pipeline_key *key) in lower_intrinsics() argument
324 nir_function_impl *entry = nir_shader_get_entrypoint(nir); in lower_intrinsics()
371 radv_lower_primitive_shading_rate(nir_shader *nir, enum amd_gfx_level gfx_level) in radv_lower_primitive_shading_rate() argument
373 nir_function_impl *impl = nir_shader_get_entrypoint(nir); in radv_lower_primitive_shading_rate()
431 if (nir->info.stage == MESA_SHADER_MESH) { in radv_lower_primitive_shading_rate()
441 if (nir->info.stage == MESA_SHADER_VERTEX) in radv_lower_primitive_shading_rate()
444 if (nir->info.stage == MESA_SHADER_VERTEX && progress) in radv_lower_primitive_shading_rate()
457 radv_force_primitive_shading_rate(nir_shader *nir, struct radv_device *device) in radv_force_primitive_shading_rate() argument
459 nir_function_impl *impl = nir_shader_get_entrypoint(nir); in radv_force_primitive_shading_rate()
489 var = nir_variable_create(nir, nir_var_shader_out, glsl_int_type(), "vrs rate"); in radv_force_primitive_shading_rate()
503 nir->info.outputs_written |= BITFIELD64_BIT(VARYING_SLOT_PRIMITIVE_SHADING_RATE); in radv_force_primitive_shading_rate()
506 if (nir->info.stage == MESA_SHADER_VERTEX) in radv_force_primitive_shading_rate()
510 if (nir->info.stage == MESA_SHADER_VERTEX && progress) in radv_force_primitive_shading_rate()
523 radv_lower_fs_intrinsics(nir_shader *nir, const struct radv_pipeline_stage *fs_stage, in radv_lower_fs_intrinsics() argument
528 nir_function_impl *impl = nir_shader_get_entrypoint(nir); in radv_lower_fs_intrinsics()
614 radv_lower_ms_workgroup_id(nir_shader *nir) in radv_lower_ms_workgroup_id() argument
616 nir_function_impl *impl = nir_shader_get_entrypoint(nir); in radv_lower_ms_workgroup_id()
684 nir_shader *nir; in radv_shader_spirv_to_nir() local
687 /* Some things such as our meta clear/blit code will give us a NIR in radv_shader_spirv_to_nir()
689 * and just use the NIR shader. We don't want to alter meta and RT in radv_shader_spirv_to_nir()
691 nir = nir_shader_clone(NULL, stage->internal_nir); in radv_shader_spirv_to_nir()
692 nir_validate_shader(nir, "in internal shader"); in radv_shader_spirv_to_nir()
694 assert(exec_list_length(&nir->functions) == 1); in radv_shader_spirv_to_nir()
790 nir = spirv_to_nir(spirv, stage->spirv.size / 4, spec_entries, num_spec_entries, stage->stage, in radv_shader_spirv_to_nir()
793 nir->info.internal |= device->app_shaders_internal; in radv_shader_spirv_to_nir()
794 assert(nir->info.stage == stage->stage); in radv_shader_spirv_to_nir()
795 nir_validate_shader(nir, "after spirv_to_nir"); in radv_shader_spirv_to_nir()
802 NIR_PASS_V(nir, nir_lower_sysvals_to_varyings, &sysvals_to_varyings); in radv_shader_spirv_to_nir()
808 NIR_PASS(_, nir, nir_lower_variable_initializers, nir_var_function_temp); in radv_shader_spirv_to_nir()
809 NIR_PASS(_, nir, nir_lower_returns); in radv_shader_spirv_to_nir()
811 NIR_PASS(progress, nir, nir_inline_functions); in radv_shader_spirv_to_nir()
813 NIR_PASS(_, nir, nir_opt_copy_prop_vars); in radv_shader_spirv_to_nir()
814 NIR_PASS(_, nir, nir_copy_prop); in radv_shader_spirv_to_nir()
816 NIR_PASS(_, nir, nir_opt_deref); in radv_shader_spirv_to_nir()
819 foreach_list_typed_safe(nir_function, func, node, &nir->functions) in radv_shader_spirv_to_nir()
826 assert(exec_list_length(&nir->functions) == 1); in radv_shader_spirv_to_nir()
831 NIR_PASS(_, nir, nir_lower_variable_initializers, nir_var_shader_out); in radv_shader_spirv_to_nir()
836 NIR_PASS(_, nir, nir_lower_variable_initializers, ~0); in radv_shader_spirv_to_nir()
841 NIR_PASS(_, nir, nir_split_var_copies); in radv_shader_spirv_to_nir()
842 NIR_PASS(_, nir, nir_split_per_member_structs); in radv_shader_spirv_to_nir()
844 if (nir->info.stage == MESA_SHADER_FRAGMENT) in radv_shader_spirv_to_nir()
845 NIR_PASS(_, nir, nir_lower_io_to_vector, nir_var_shader_out); in radv_shader_spirv_to_nir()
846 if (nir->info.stage == MESA_SHADER_FRAGMENT) in radv_shader_spirv_to_nir()
847 NIR_PASS(_, nir, nir_lower_input_attachments, in radv_shader_spirv_to_nir()
856 NIR_PASS(_, nir, nir_remove_dead_variables, in radv_shader_spirv_to_nir()
863 NIR_PASS(_, nir, nir_lower_global_vars_to_local); in radv_shader_spirv_to_nir()
864 NIR_PASS(_, nir, nir_lower_vars_to_ssa); in radv_shader_spirv_to_nir()
866 NIR_PASS(_, nir, nir_propagate_invariant, key->invariant_geom); in radv_shader_spirv_to_nir()
868 NIR_PASS(_, nir, nir_lower_clip_cull_distance_arrays); in radv_shader_spirv_to_nir()
870 if (nir->info.stage == MESA_SHADER_VERTEX || in radv_shader_spirv_to_nir()
871 nir->info.stage == MESA_SHADER_TESS_EVAL || in radv_shader_spirv_to_nir()
872 nir->info.stage == MESA_SHADER_GEOMETRY) in radv_shader_spirv_to_nir()
873 NIR_PASS_V(nir, nir_shader_gather_xfb_info); in radv_shader_spirv_to_nir()
875 NIR_PASS(_, nir, nir_lower_discard_or_demote, key->ps.lower_discard_to_demote); in radv_shader_spirv_to_nir()
877 nir_lower_doubles_options lower_doubles = nir->options->lower_doubles_options; in radv_shader_spirv_to_nir()
887 NIR_PASS(_, nir, nir_lower_doubles, NULL, lower_doubles); in radv_shader_spirv_to_nir()
889 NIR_PASS(_, nir, nir_shader_lower_instructions, &is_sincos, &lower_sincos, NULL); in radv_shader_spirv_to_nir()
892 NIR_PASS(_, nir, nir_lower_system_values); in radv_shader_spirv_to_nir()
897 .lower_cs_local_id_to_index = nir->info.stage == MESA_SHADER_MESH, in radv_shader_spirv_to_nir()
898 .lower_local_invocation_index = nir->info.stage == MESA_SHADER_COMPUTE && in radv_shader_spirv_to_nir()
899 ((nir->info.workgroup_size[0] == 1) + in radv_shader_spirv_to_nir()
900 (nir->info.workgroup_size[1] == 1) + in radv_shader_spirv_to_nir()
901 (nir->info.workgroup_size[2] == 1)) == 2, in radv_shader_spirv_to_nir()
903 NIR_PASS(_, nir, nir_lower_compute_system_values, &csv_options); in radv_shader_spirv_to_nir()
905 if (nir->info.stage == MESA_SHADER_MESH) { in radv_shader_spirv_to_nir()
907 NIR_PASS(_, nir, radv_lower_ms_workgroup_id); in radv_shader_spirv_to_nir()
915 NIR_PASS(_, nir, nir_lower_compute_system_values, &o); in radv_shader_spirv_to_nir()
919 nir->info.separate_shader = true; in radv_shader_spirv_to_nir()
921 nir_shader_gather_info(nir, nir_shader_get_entrypoint(nir)); in radv_shader_spirv_to_nir()
923 if (nir->info.ray_queries > 0) { in radv_shader_spirv_to_nir()
924 NIR_PASS(_, nir, nir_opt_ray_queries); in radv_shader_spirv_to_nir()
925 NIR_PASS(_, nir, radv_nir_lower_ray_queries, device); in radv_shader_spirv_to_nir()
939 NIR_PASS(_, nir, nir_lower_tex, &tex_options); in radv_shader_spirv_to_nir()
945 NIR_PASS(_, nir, nir_lower_image, &image_options); in radv_shader_spirv_to_nir()
947 NIR_PASS(_, nir, nir_lower_vars_to_ssa); in radv_shader_spirv_to_nir()
949 if (nir->info.stage == MESA_SHADER_VERTEX || nir->info.stage == MESA_SHADER_GEOMETRY || in radv_shader_spirv_to_nir()
950 nir->info.stage == MESA_SHADER_FRAGMENT) { in radv_shader_spirv_to_nir()
951 NIR_PASS_V(nir, nir_lower_io_to_temporaries, nir_shader_get_entrypoint(nir), true, true); in radv_shader_spirv_to_nir()
952 } else if (nir->info.stage == MESA_SHADER_TESS_EVAL) { in radv_shader_spirv_to_nir()
953 NIR_PASS_V(nir, nir_lower_io_to_temporaries, nir_shader_get_entrypoint(nir), true, false); in radv_shader_spirv_to_nir()
956 NIR_PASS(_, nir, nir_split_var_copies); in radv_shader_spirv_to_nir()
958 NIR_PASS(_, nir, nir_lower_global_vars_to_local); in radv_shader_spirv_to_nir()
959 NIR_PASS(_, nir, nir_remove_dead_variables, nir_var_function_temp, NULL); in radv_shader_spirv_to_nir()
961 NIR_PASS(_, nir, nir_lower_subgroups, in radv_shader_spirv_to_nir()
976 NIR_PASS(_, nir, nir_lower_load_const_to_scalar); in radv_shader_spirv_to_nir()
977 NIR_PASS(_, nir, nir_opt_shrink_stores, !device->instance->disable_shrink_image_store); in radv_shader_spirv_to_nir()
980 radv_optimize_nir(nir, false, true); in radv_shader_spirv_to_nir()
985 NIR_PASS(_, nir, nir_lower_var_copies); in radv_shader_spirv_to_nir()
987 unsigned lower_flrp = (nir->options->lower_flrp16 ? 16 : 0) | in radv_shader_spirv_to_nir()
988 (nir->options->lower_flrp32 ? 32 : 0) | in radv_shader_spirv_to_nir()
989 (nir->options->lower_flrp64 ? 64 : 0); in radv_shader_spirv_to_nir()
992 NIR_PASS(progress, nir, nir_lower_flrp, lower_flrp, false /* always precise */); in radv_shader_spirv_to_nir()
994 NIR_PASS(_, nir, nir_opt_constant_folding); in radv_shader_spirv_to_nir()
1001 NIR_PASS(_, nir, nir_opt_access, &opt_access_options); in radv_shader_spirv_to_nir()
1003 NIR_PASS(_, nir, nir_lower_explicit_io, nir_var_mem_push_const, nir_address_format_32bit_offset); in radv_shader_spirv_to_nir()
1005 NIR_PASS(_, nir, nir_lower_explicit_io, nir_var_mem_ubo | nir_var_mem_ssbo, in radv_shader_spirv_to_nir()
1008 NIR_PASS(_, nir, lower_intrinsics, key); in radv_shader_spirv_to_nir()
1011 if (nir->info.stage == MESA_SHADER_COMPUTE || in radv_shader_spirv_to_nir()
1012 nir->info.stage == MESA_SHADER_TASK || in radv_shader_spirv_to_nir()
1013 nir->info.stage == MESA_SHADER_MESH) { in radv_shader_spirv_to_nir()
1016 if (nir->info.stage == MESA_SHADER_TASK || in radv_shader_spirv_to_nir()
1017 nir->info.stage == MESA_SHADER_MESH) in radv_shader_spirv_to_nir()
1020 if (!nir->info.shared_memory_explicit_layout) { in radv_shader_spirv_to_nir()
1021 NIR_PASS(_, nir, nir_lower_vars_to_explicit_types, var_modes, shared_var_info); in radv_shader_spirv_to_nir()
1023 NIR_PASS(_, nir, nir_lower_explicit_io, var_modes, nir_address_format_32bit_offset); in radv_shader_spirv_to_nir()
1025 if (nir->info.zero_initialize_shared_memory && nir->info.shared_size > 0) { in radv_shader_spirv_to_nir()
1027 const unsigned shared_size = ALIGN(nir->info.shared_size, chunk_size); in radv_shader_spirv_to_nir()
1028 NIR_PASS(_, nir, nir_zero_initialize_shared_memory, shared_size, chunk_size); in radv_shader_spirv_to_nir()
1032 NIR_PASS(_, nir, nir_lower_explicit_io, nir_var_mem_global | nir_var_mem_constant, in radv_shader_spirv_to_nir()
1039 NIR_PASS(_, nir, nir_opt_large_constants, glsl_get_natural_size_align_bytes, 16); in radv_shader_spirv_to_nir()
1042 if ((nir->info.stage == MESA_SHADER_VERTEX || in radv_shader_spirv_to_nir()
1043 nir->info.stage == MESA_SHADER_GEOMETRY || in radv_shader_spirv_to_nir()
1044 nir->info.stage == MESA_SHADER_MESH) && in radv_shader_spirv_to_nir()
1045 nir->info.outputs_written & BITFIELD64_BIT(VARYING_SLOT_PRIMITIVE_SHADING_RATE)) { in radv_shader_spirv_to_nir()
1047 NIR_PASS(_, nir, radv_lower_primitive_shading_rate, in radv_shader_spirv_to_nir()
1056 if (ac_nir_lower_indirect_derefs(nir, device->physical_device->rad_info.gfx_level) && in radv_shader_spirv_to_nir()
1057 !key->optimisations_disabled && nir->info.stage != MESA_SHADER_COMPUTE) { in radv_shader_spirv_to_nir()
1059 radv_optimize_nir(nir, false, false); in radv_shader_spirv_to_nir()
1063 return nir; in radv_shader_spirv_to_nir()
1073 find_layer_in_var(nir_shader *nir) in find_layer_in_var() argument
1075 nir_variable *var = nir_find_variable_with_location(nir, nir_var_shader_in, VARYING_SLOT_LAYER); in find_layer_in_var()
1079 var = nir_variable_create(nir, nir_var_shader_in, glsl_int_type(), "layer id"); in find_layer_in_var()
1095 lower_view_index(nir_shader *nir, bool per_primitive) in lower_view_index() argument
1098 nir_function_impl *entry = nir_shader_get_entrypoint(nir); in lower_view_index()
1113 layer = find_layer_in_var(nir); in lower_view_index()
1121 nir->info.inputs_read |= VARYING_BIT_LAYER; in lower_view_index()
1123 nir->info.per_primitive_inputs |= VARYING_BIT_LAYER; in lower_view_index()
1139 radv_lower_io(struct radv_device *device, nir_shader *nir, bool is_mesh_shading) in radv_lower_io() argument
1141 if (nir->info.stage == MESA_SHADER_COMPUTE) in radv_lower_io()
1144 if (nir->info.stage == MESA_SHADER_FRAGMENT) { in radv_lower_io()
1145 NIR_PASS(_, nir, lower_view_index, is_mesh_shading); in radv_lower_io()
1146 nir_assign_io_var_locations(nir, nir_var_shader_in, &nir->num_inputs, MESA_SHADER_FRAGMENT); in radv_lower_io()
1149 NIR_PASS(_, nir, nir_lower_io, nir_var_shader_in | nir_var_shader_out, type_size_vec4, in radv_lower_io()
1153 NIR_PASS(_, nir, nir_opt_constant_folding); in radv_lower_io()
1155 NIR_PASS(_, nir, nir_io_add_const_offset_to_base, nir_var_shader_in | nir_var_shader_out); in radv_lower_io()
1163 nir_shader *nir = stage->nir; in radv_lower_io_to_mem() local
1165 if (nir->info.stage == MESA_SHADER_VERTEX) { in radv_lower_io_to_mem()
1167 NIR_PASS_V(nir, ac_nir_lower_ls_outputs_to_mem, NULL, info->vs.tcs_in_out_eq, in radv_lower_io_to_mem()
1171 NIR_PASS_V(nir, ac_nir_lower_es_outputs_to_mem, NULL, in radv_lower_io_to_mem()
1176 } else if (nir->info.stage == MESA_SHADER_TESS_CTRL) { in radv_lower_io_to_mem()
1177 NIR_PASS_V(nir, ac_nir_lower_hs_inputs_to_mem, NULL, info->vs.tcs_in_out_eq); in radv_lower_io_to_mem()
1178 NIR_PASS_V(nir, ac_nir_lower_hs_outputs_to_mem, NULL, in radv_lower_io_to_mem()
1186 } else if (nir->info.stage == MESA_SHADER_TESS_EVAL) { in radv_lower_io_to_mem()
1187 NIR_PASS_V(nir, ac_nir_lower_tes_inputs_to_mem, NULL); in radv_lower_io_to_mem()
1190 NIR_PASS_V(nir, ac_nir_lower_es_outputs_to_mem, NULL, in radv_lower_io_to_mem()
1196 } else if (nir->info.stage == MESA_SHADER_GEOMETRY) { in radv_lower_io_to_mem()
1197 NIR_PASS_V(nir, ac_nir_lower_gs_inputs_to_mem, NULL, in radv_lower_io_to_mem()
1200 } else if (nir->info.stage == MESA_SHADER_TASK) { in radv_lower_io_to_mem()
1201 ac_nir_apply_first_task_to_task_shader(nir); in radv_lower_io_to_mem()
1202 ac_nir_lower_task_outputs_to_mem(nir, AC_TASK_PAYLOAD_ENTRY_BYTES, in radv_lower_io_to_mem()
1205 } else if (nir->info.stage == MESA_SHADER_MESH) { in radv_lower_io_to_mem()
1206 ac_nir_lower_mesh_inputs_to_mem(nir, AC_TASK_PAYLOAD_ENTRY_BYTES, in radv_lower_io_to_mem()
1215 radv_consider_culling(const struct radv_physical_device *pdevice, struct nir_shader *nir, uint64_t … in radv_consider_culling() argument
1219 if (is_meta_shader(nir)) in radv_consider_culling()
1223 if (nir->info.outputs_written & (VARYING_BIT_VIEWPORT | VARYING_BIT_VIEWPORT_MASK)) in radv_consider_culling()
1259 if (nir->info.writes_memory) in radv_consider_culling()
1265 if (BITSET_TEST(nir->info.system_values_read, SYSTEM_VALUE_SUBGROUP_INVOCATION)) in radv_consider_culling()
1275 nir_shader *nir = ngg_stage->nir; in radv_lower_ngg() local
1277 assert(nir->info.stage == MESA_SHADER_VERTEX || in radv_lower_ngg()
1278 nir->info.stage == MESA_SHADER_TESS_EVAL || in radv_lower_ngg()
1279 nir->info.stage == MESA_SHADER_GEOMETRY || in radv_lower_ngg()
1280 nir->info.stage == MESA_SHADER_MESH); in radv_lower_ngg()
1286 if (nir->info.stage == MESA_SHADER_TESS_EVAL) { in radv_lower_ngg()
1287 if (nir->info.tess.point_mode) in radv_lower_ngg()
1289 else if (nir->info.tess._primitive_mode == TESS_PRIMITIVE_ISOLINES) in radv_lower_ngg()
1294 BITSET_SET(nir->info.system_values_read, SYSTEM_VALUE_PRIMITIVE_ID); in radv_lower_ngg()
1296 } else if (nir->info.stage == MESA_SHADER_VERTEX) { in radv_lower_ngg()
1302 BITSET_SET(nir->info.system_values_read, SYSTEM_VALUE_INSTANCE_ID); in radv_lower_ngg()
1304 } else if (nir->info.stage == MESA_SHADER_GEOMETRY) { in radv_lower_ngg()
1305 num_vertices_per_prim = nir->info.gs.vertices_in; in radv_lower_ngg()
1306 } else if (nir->info.stage == MESA_SHADER_MESH) { in radv_lower_ngg()
1307 if (nir->info.mesh.primitive_type == SHADER_PRIM_POINTS) in radv_lower_ngg()
1309 else if (nir->info.mesh.primitive_type == SHADER_PRIM_LINES) in radv_lower_ngg()
1312 assert(nir->info.mesh.primitive_type == SHADER_PRIM_TRIANGLES); in radv_lower_ngg()
1320 if (nir->info.stage == MESA_SHADER_VERTEX || in radv_lower_ngg()
1321 nir->info.stage == MESA_SHADER_TESS_EVAL) { in radv_lower_ngg()
1327 radv_optimize_nir_algebraic(nir, false); in radv_lower_ngg()
1329 if (nir->info.stage == MESA_SHADER_VERTEX) { in radv_lower_ngg()
1335 NIR_PASS_V(nir, ac_nir_lower_ngg_nogs, in radv_lower_ngg()
1344 ngg_stage->info.ngg_info.esgs_ring_size = nir->info.shared_size; in radv_lower_ngg()
1345 } else if (nir->info.stage == MESA_SHADER_GEOMETRY) { in radv_lower_ngg()
1347 NIR_PASS_V(nir, ac_nir_lower_ngg_gs, info->wave_size, info->workgroup_size, in radv_lower_ngg()
1350 } else if (nir->info.stage == MESA_SHADER_MESH) { in radv_lower_ngg()
1352 …NIR_PASS_V(nir, ac_nir_lower_ngg_ms, &scratch_ring, info->wave_size, pl_key->has_multiview_view_in… in radv_lower_ngg()