Home
last modified time | relevance | path

Searched refs:GFX9 (Results 1 – 25 of 154) sorted by relevance

1234567

/third_party/skia/third_party/externals/swiftshader/third_party/llvm-10.0/llvm/lib/Target/AMDGPU/
DAMDGPUSubtarget.h58 GFX9 = 7, enumerator
534 return getGeneration() >= AMDGPUSubtarget::GFX9; in hasMed3_16()
538 return getGeneration() >= AMDGPUSubtarget::GFX9; in hasMin3Max3_16()
635 return getGeneration() >= AMDGPUSubtarget::GFX9; in supportsMinMaxDenormModes()
661 return getGeneration() < AMDGPUSubtarget::GFX9; in privateMemoryResourceIsRangeChecked()
732 return getGeneration() > GFX9; in hasFlatLgkmVMemCountInOrder()
736 return getGeneration() >= GFX9; in hasD16LoadStore()
750 return getGeneration() < GFX9; in ldsRequiresM0Init()
759 return getGeneration() >= GFX9; in hasGWSAutoReplay()
1018 return getGeneration() == AMDGPUSubtarget::GFX9; in hasSMovFedHazard()
[all …]
/third_party/mesa3d/src/amd/common/
Dac_surface_test_common.h38 info->chip_class = GFX9; in init_vega10()
53 info->chip_class = GFX9; in init_vega20()
69 info->chip_class = GFX9; in init_raven()
84 info->chip_class = GFX9; in init_raven2()
205 case GFX9: in get_radeon_info()
Dac_shader_util.c105 S_028A40_ONCHIP(chip_class >= GFX9 ? 1 : 0); in ac_vgt_gs_mode()
228 if (chip_class == GFX9) in ac_get_sampler_dim()
258 else if (sdim == GLSL_SAMPLER_DIM_2D && !is_array && chip_class == GFX9) { in ac_get_image_dim()
533 bool merged_shaders = chip_class >= GFX9; in ac_compute_lshs_workgroup_size()
Dac_gpu_info.c315 if (info->chip_class < GFX9) in has_tmz_support()
640 info->chip_class = GFX9; in ac_query_gpu_info()
680 info->has_l2_uncached = info->chip_class >= GFX9; in ac_query_gpu_info()
800 if (info->chip_class >= GFX9) { in ac_query_gpu_info()
857 info->has_rbplus = info->family == CHIP_STONEY || info->chip_class >= GFX9; in ac_query_gpu_info()
868 info->chip_class >= GFX8 && info->chip_class <= GFX9 && info->max_se >= 2; in ac_query_gpu_info()
871 info->has_packed_math_16bit = info->chip_class >= GFX9; in ac_query_gpu_info()
882 info->chip_class >= GFX9 || (info->chip_class >= GFX8 && info->me_fw_feature >= 41); in ac_query_gpu_info()
888 info->has_tc_compat_zrange_bug = info->chip_class >= GFX8 && info->chip_class <= GFX9; in ac_query_gpu_info()
932 (info->chip_class == GFX9 && in ac_query_gpu_info()
[all …]
Dac_nir_lower_esgs_io_to_mem.c196 nir_ssa_def *vertex_offset = st->chip_class >= GFX9 in gs_per_vertex_input_offset()
200 unsigned base_stride = st->chip_class >= GFX9 ? 1 : 64 /* Wave size on GFX6-8 */; in gs_per_vertex_input_offset()
215 if (st->chip_class >= GFX9) in lower_gs_per_vertex_input_load()
Dac_shadowed_regs.c838 else if (chip_class == GFX9) in ac_get_reg_ranges()
846 else if (chip_class == GFX9) in ac_get_reg_ranges()
854 else if (chip_class == GFX9) in ac_get_reg_ranges()
862 else if (chip_class == GFX9) in ac_get_reg_ranges()
2942 } else if (info->chip_class == GFX9) { in ac_emulate_clear_state()
Dac_surface_meta_address_test.c264 if (info->chip_class == GFX9) { in one_dcc_address_test()
287 if (info->chip_class == GFX9) { in one_dcc_address_test()
321 unsigned swizzle_mode = info->chip_class == GFX9 ? ADDR_SW_64KB_S_X : ADDR_SW_64KB_R_X; in run_dcc_address_test()
352 …for (unsigned samples = 1; samples <= (info->chip_class == GFX9 ? max_samples : 1); samples *= 2) { in run_dcc_address_test()
606 if (info->chip_class == GFX9) { in one_cmask_address_test()
640 unsigned swizzle_mode = info->chip_class == GFX9 ? ADDR_SW_64KB_S_X : ADDR_SW_64KB_Z_X; in run_cmask_address_test()
Dac_nir.c40 bool llvm_has_working_vgpr_indexing = chip_class != GFX9; in ac_nir_lower_indirect_derefs()
/third_party/mesa3d/src/amd/compiler/
Daco_assembler.cpp55 else if (chip_class <= GFX9) in asm_context()
150 if (opcode >= 55 && ctx.chip_class <= GFX9) { in emit_instruction()
151 assert(ctx.chip_class == GFX9 && opcode < 60); in emit_instruction()
209 if (ctx.chip_class <= GFX9) { in emit_instruction()
222 if (ctx.chip_class <= GFX9) { in emit_instruction()
226 if (ctx.chip_class == GFX9) { in emit_instruction()
248 if (ctx.chip_class <= GFX9) { in emit_instruction()
263 assert(ctx.chip_class >= GFX9); /* GFX8 and below don't support specifying a constant in emit_instruction()
310 if (ctx.chip_class == GFX8 || ctx.chip_class == GFX9) { in emit_instruction()
333 if (ctx.chip_class == GFX8 || ctx.chip_class == GFX9) { in emit_instruction()
[all …]
Daco_print_asm.cpp81 case GFX9: in to_clrx_device_name()
156 if (((chip == GFX8 || chip == GFX9) && (binary[pos] & 0xffff8000) == 0xd28a0000) || in disasm_instr()
174 ((chip >= GFX9 && (binary[pos] & 0xffff8000) == 0xd1348000) || /* v_add_u32_e64 + clamp */ in disasm_instr()
176 (chip <= GFX9 && (binary[pos] & 0xffff8000) == 0xd1268000) || /* v_add_u16_e64 + clamp */ in disasm_instr()
178 (chip == GFX9 && (binary[pos] & 0xffff8000) == 0xd1ff8000)) /* v_add3_u32 + clamp */) { in disasm_instr()
Daco_ir.cpp81 case GFX9: program->family = CHIP_VEGA10; break; in init_program()
147 program->dev.has_fast_fma32 = program->chip_class >= GFX9; in init_program()
202 if (vop3.omod && chip < GFX9) in can_use_SDWA()
212 if (chip < GFX9 && !instr->operands[i].isOfType(RegType::vgpr)) in can_use_SDWA()
223 if (chip < GFX9 && !instr->operands[0].isOfType(RegType::vgpr)) in can_use_SDWA()
372 if ((high || idx == -1) && chip < GFX9) in can_use_opsel()
415 if (chip < GFX9) in instr_is_16bit()
430 case aco_opcode::v_madmk_f16: return chip >= GFX9; in instr_is_16bit()
741 if (chip >= GFX9) in wait_imm()
763 case GFX9: in pack()
[all …]
/third_party/mesa3d/src/amd/vulkan/winsys/null/
Dradv_null_winsys.c91 info->chip_class = GFX9; in radv_null_winsys_query_info()
134 info->has_packed_math_16bit = info->chip_class >= GFX9; in radv_null_winsys_query_info()
143 info->address32_hi = info->chip_class >= GFX9 ? 0xffff8000u : 0x0; in radv_null_winsys_query_info()
145 info->has_rbplus = info->family == CHIP_STONEY || info->chip_class >= GFX9; in radv_null_winsys_query_info()
/third_party/mesa3d/src/intel/compiler/
Dbrw_gfx_ver_enum.h35 GFX9 = (1 << 7), enumerator
58 case 90: return GFX9; in gfx_ver_from_devinfo()
/third_party/mesa3d/src/amd/vulkan/
Dsi_cmd_buffer.c106 if (device->physical_device->rad_info.chip_class >= GFX9) { in si_emit_compute()
217 if (physical_device->rad_info.chip_class <= GFX9) in si_emit_graphics()
277 } else if (physical_device->rad_info.chip_class == GFX9) { in si_emit_graphics()
301 } else if (device->physical_device->rad_info.chip_class == GFX9) { in si_emit_graphics()
337 if (physical_device->rad_info.chip_class >= GFX9) { in si_emit_graphics()
444 if (physical_device->rad_info.chip_class >= GFX9) { in si_emit_graphics()
475 if (physical_device->rad_info.chip_class >= GFX9) { in si_emit_graphics()
846 unsigned is_gfx8_mec = is_mec && chip_class < GFX9; in si_cs_emit_write_event_eop()
854 if (chip_class >= GFX9 || is_gfx8_mec) { in si_cs_emit_write_event_eop()
859 if (chip_class == GFX9 && !is_mec) { in si_cs_emit_write_event_eop()
[all …]
Dradv_image.c93 if (device->physical_device->rad_info.chip_class < GFX9) { in radv_use_tc_compat_htile_for_image()
117 if (device->physical_device->rad_info.chip_class >= GFX9) in radv_surface_has_scanout()
283 device->physical_device->rad_info.chip_class == GFX9) in radv_use_dcc_for_image_early()
422 if (device->physical_device->rad_info.chip_class >= GFX9) { in radv_patch_surface_from_metadata()
586 if (device->physical_device->rad_info.chip_class >= GFX9 && in radv_get_surface_flags()
731 if (chip_class >= GFX9) { in si_set_mutable_tex_desc_fields()
740 if (chip_class >= GFX9 || base_level_info->mode == RADEON_SURF_MODE_2D) in si_set_mutable_tex_desc_fields()
762 if (chip_class <= GFX9) in si_set_mutable_tex_desc_fields()
795 } else if (chip_class == GFX9) { in si_set_mutable_tex_desc_fields()
924 is_storage_image, device->physical_device->rad_info.chip_class == GFX9); in gfx10_make_texture_descriptor()
[all …]
Dradv_sqtt.c134 if (device->physical_device->rad_info.chip_class < GFX9) { in radv_emit_thread_trace_start()
153 if (device->physical_device->rad_info.chip_class == GFX9) { in radv_emit_thread_trace_start()
165 if (device->physical_device->rad_info.chip_class == GFX9) { in radv_emit_thread_trace_start()
215 } else if (device->physical_device->rad_info.chip_class == GFX9) { in radv_copy_thread_trace_info_regs()
341 if (device->physical_device->rad_info.chip_class >= GFX9) { in radv_emit_spi_config_cntl()
Dradv_cs.h167 if (pdevice->rad_info.chip_class < GFX9 || in radeon_set_uconfig_reg_idx()
168 (pdevice->rad_info.chip_class == GFX9 && pdevice->rad_info.me_fw_version < 26)) in radeon_set_uconfig_reg_idx()
/third_party/mesa3d/src/gallium/drivers/radeonsi/
Dsi_build_pm4.h147 if ((chip_class) < GFX9 || \
148 ((chip_class) == GFX9 && (screen)->info.me_fw_version < 26)) \
294 } else if (chip_class == GFX9) { in si_get_user_data_base()
312 if (chip_class == GFX9) { in si_get_user_data_base()
337 if (chip_class == GFX9) { in si_get_user_data_base()
Dsi_shader_llvm_gs.c58 if (ctx->screen->info.chip_class >= GFX9) { in si_llvm_load_input_gs()
145 if (ctx->screen->info.chip_class >= GFX9 && info->num_outputs) { in si_llvm_emit_es_epilogue()
175 if (ctx->screen->info.chip_class >= GFX9) { in si_llvm_emit_es_epilogue()
188 if (ctx->screen->info.chip_class >= GFX9) in si_llvm_emit_es_epilogue()
194 if (ctx->screen->info.chip_class >= GFX9) in si_get_gs_wave_id()
212 if (ctx->screen->info.chip_class >= GFX9) in emit_gs_epilogue()
561 if (ctx->screen->info.chip_class >= GFX9) { in si_llvm_build_gs_prolog()
613 if (ctx->screen->info.chip_class >= GFX9) { in si_llvm_build_gs_prolog()
633 if (ctx->screen->info.chip_class >= GFX9) { in si_llvm_build_gs_prolog()
Dsi_state_draw.cpp40 #define GFX(name) name##GFX9
221 if (GFX_VERSION >= GFX9 && HAS_TESS) in si_update_shaders()
223 else if (GFX_VERSION >= GFX9 && HAS_GS) in si_update_shaders()
260 if ((GFX_VERSION >= GFX10_3 || (GFX_VERSION >= GFX9 && sctx->screen->info.rbplus_allowed)) && in si_update_shaders()
417 if (GFX_VERSION >= GFX9) { in si_prefetch_shaders()
524 if (sctx->chip_class >= GFX9) { in si_emit_derived_tess_state()
713 if (sctx->chip_class >= GFX9) { in si_emit_derived_tess_state()
908 S_030960_EN_INST_OPT_BASIC(sscreen->info.chip_class >= GFX9) | in si_get_init_multi_vgt_param()
909 S_030960_EN_INST_OPT_ADV(sscreen->info.chip_class >= GFX9); in si_get_init_multi_vgt_param()
1085 if (GFX_VERSION == GFX9) in si_emit_rasterizer_prim_state()
[all …]
Dsi_fence.c80 if (ctx->chip_class >= GFX9 || (compute_ib && ctx->chip_class >= GFX7)) { in si_cp_release_mem()
88 if (ctx->chip_class == GFX9 && !compute_ib && query_type != PIPE_QUERY_OCCLUSION_COUNTER && in si_cp_release_mem()
104 radeon_emit(PKT3(PKT3_RELEASE_MEM, ctx->chip_class >= GFX9 ? 6 : 5, 0)); in si_cp_release_mem()
111 if (ctx->chip_class >= GFX9) in si_cp_release_mem()
Dsi_cp_dma.c46 sctx->chip_class >= GFX9 ? S_415_BYTE_COUNT_GFX9(~0u) : S_415_BYTE_COUNT_GFX6(~0u); in cp_dma_max_byte_count()
65 if (sctx->chip_class >= GFX9) in si_emit_cp_dma()
78 if (sctx->chip_class >= GFX9 && !(flags & CP_DMA_CLEAR) && src_va == dst_va) { in si_emit_cp_dma()
412 if (sctx->chip_class >= GFX9) { in si_cp_dma_prefetch()
/third_party/mesa3d/docs/relnotes/
D18.1.2.rst38 - radv: Consolidate GFX9 merged shader lookup logic
39 - radv: Handle GFX9 merged shaders in radv_flush_constants()
123 - radv: fix missing ZRANGE_PRECISION(1) for GFX9+
D17.2.4.rst45 - cherry-ignore: radv: Don't use vgpr indexing for outputs on GFX9.
46 - cherry-ignore: radv: Disallow indirect outputs for GS on GFX9 as
/third_party/mesa3d/src/amd/compiler/tests/
Dtest_optimizer.cpp29 for (unsigned i = GFX9; i <= GFX10; i++) {
90 if (!setup_cs("v1 v1", GFX9))
273 for (unsigned i = GFX9; i <= GFX10; i++) {
475 if (!setup_cs("v1 v1 v1", GFX9, CHIP_UNKNOWN, cfg.name))
555 if (!setup_cs("v1 v1 v2 v1", GFX9))
688 if (!setup_cs("v1 v1 v1", GFX9))
715 for (unsigned i = GFX9; i <= GFX10; i++) {
738 for (unsigned i = GFX8; i <= GFX9; i++) {
867 for (unsigned i = GFX8; i <= GFX9; i++) {
892 …bool can_propagate = cfg.src == aco_opcode::v_rcp_f32 || (i >= GFX9 && cfg.src == aco_opcode::v_mi…
[all …]

1234567