Home
last modified time | relevance | path

Searched refs:GFX9 (Results 1 – 25 of 164) 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 …]
DAMDGPU.td248 "Additional instructions for GFX9+"
260 "Instructions shared in GFX7, GFX8, GFX9"
657 def FeatureGFX9 : GCNSubtargetFeatureGeneration<"GFX9",
1001 "Subtarget->getGeneration() == AMDGPUSubtarget::GFX9">,
1008 "Subtarget->getGeneration() == AMDGPUSubtarget::GFX9">,
1024 Predicate<"Subtarget->getGeneration() >= AMDGPUSubtarget::GFX9">,
1028 "Subtarget->getGeneration() == AMDGPUSubtarget::GFX9">,
1033 "Subtarget->getGeneration() == AMDGPUSubtarget::GFX9">,
1064 def HasDSAddTid : Predicate<"Subtarget->getGeneration() >= AMDGPUSubtarget::GFX9">,
/third_party/mesa3d/src/amd/common/
Dac_surface_test_common.h38 info->gfx_level = GFX9; in init_vega10()
53 info->gfx_level = GFX9; in init_vega20()
69 info->gfx_level = GFX9; in init_raven()
84 info->gfx_level = GFX9; in init_raven2()
196 case GFX9: in get_radeon_info()
Dac_nir_lower_esgs_io_to_mem.c193 if (st->gfx_level < GFX9) { in gs_get_vertex_offset()
198 assert(st->gfx_level == GFX9); in gs_get_vertex_offset()
258 nir_ssa_def *vertex_offset = st->gfx_level >= GFX9 in gs_per_vertex_input_offset()
262 unsigned base_stride = st->gfx_level >= GFX9 ? 1 : 64 /* Wave size on GFX6-8 */; in gs_per_vertex_input_offset()
277 if (st->gfx_level >= GFX9) in lower_gs_per_vertex_input_load()
Dac_gpu_info.c419 if (info->gfx_level < GFX9) in has_tmz_support()
856 info->gfx_level = GFX9; in ac_query_gpu_info()
897 info->has_l2_uncached = info->gfx_level >= GFX9; in ac_query_gpu_info()
1015 if (info->gfx_level >= GFX9) { in ac_query_gpu_info()
1059 info->has_rbplus = info->family == CHIP_STONEY || info->gfx_level >= GFX9; in ac_query_gpu_info()
1070 info->gfx_level >= GFX8 && info->gfx_level <= GFX9 && info->max_se >= 2; in ac_query_gpu_info()
1073 info->has_packed_math_16bit = info->gfx_level >= GFX9; in ac_query_gpu_info()
1084 info->gfx_level >= GFX9 || (info->gfx_level >= GFX8 && info->me_fw_feature >= 41); in ac_query_gpu_info()
1090 info->has_tc_compat_zrange_bug = info->gfx_level >= GFX8 && info->gfx_level <= GFX9; in ac_query_gpu_info()
1134 (info->gfx_level == GFX9 && in ac_query_gpu_info()
[all …]
Dac_surface.c202 if (info->gfx_level < GFX9) in ac_is_modifier_supported()
209 if (info->gfx_level < GFX9 && util_format_get_num_planes(format) > 1) in ac_is_modifier_supported()
214 case GFX9: in ac_is_modifier_supported()
268 case GFX9: { in ac_get_supported_modifiers()
1525 if (info->gfx_level <= GFX9) { in is_dcc_supported_by_L2()
1612 case GFX9: in is_dcc_supported_by_DCN()
1676 if (info->gfx_level == GFX9) { in ac_copy_cmask_equation()
1917 if (info->gfx_level == GFX9) in gfx9_compute_miptree()
1920 if (info->gfx_level == GFX9) in gfx9_compute_miptree()
2003 if (info->gfx_level == GFX9) in gfx9_compute_miptree()
[all …]
Dac_surface_meta_address_test.c264 if (info->gfx_level == GFX9) { in one_dcc_address_test()
287 if (info->gfx_level == GFX9) { in one_dcc_address_test()
325 case GFX9: in run_dcc_address_test()
369 … for (unsigned samples = 1; samples <= (info->gfx_level == GFX9 ? max_samples : 1); samples *= 2) { in run_dcc_address_test()
505 case GFX9: in run_htile_address_test()
641 if (info->gfx_level == GFX9) { in one_cmask_address_test()
675 unsigned swizzle_mode = info->gfx_level == GFX9 ? ADDR_SW_64KB_S_X : ADDR_SW_64KB_Z_X; in run_cmask_address_test()
Dac_shader_util.c111 S_028A40_ONCHIP(gfx_level >= GFX9 ? 1 : 0); in ac_vgt_gs_mode()
429 if (gfx_level == GFX9) in ac_get_sampler_dim()
459 else if (sdim == GLSL_SAMPLER_DIM_2D && !is_array && gfx_level == GFX9) { in ac_get_image_dim()
740 bool merged_shaders = gfx_level >= GFX9; in ac_compute_lshs_workgroup_size()
/third_party/mesa3d/src/amd/compiler/
Daco_assembler.cpp55 else if (gfx_level <= GFX9) in asm_context()
207 if (ctx.gfx_level <= GFX9) { in emit_instruction()
220 if (ctx.gfx_level <= GFX9) { in emit_instruction()
224 if (ctx.gfx_level == GFX9) { in emit_instruction()
246 if (ctx.gfx_level <= GFX9) { in emit_instruction()
261 assert(ctx.gfx_level >= GFX9); /* GFX8 and below don't support specifying a constant in emit_instruction()
308 if (ctx.gfx_level == GFX8 || ctx.gfx_level == GFX9) { in emit_instruction()
331 if (ctx.gfx_level == GFX8 || ctx.gfx_level == GFX9) { in emit_instruction()
353 if (ctx.gfx_level == GFX8 || ctx.gfx_level == GFX9) { in emit_instruction()
389 if (ctx.gfx_level == GFX8 || ctx.gfx_level == GFX9) { in emit_instruction()
[all …]
Daco_ir.cpp81 case GFX9: program->family = CHIP_VEGA10; break; in init_program()
147 program->dev.has_fast_fma32 = program->gfx_level >= GFX9; in init_program()
164 } else if (program->gfx_level == GFX9) { in init_program()
220 if (vop3.omod && gfx_level < GFX9) in can_use_SDWA()
230 if (gfx_level < GFX9 && !instr->operands[i].isOfType(RegType::vgpr)) in can_use_SDWA()
241 if (gfx_level < GFX9 && !instr->operands[0].isOfType(RegType::vgpr)) in can_use_SDWA()
406 if (gfx_level < GFX9) in can_use_opsel()
449 if (gfx_level < GFX9) in instr_is_16bit()
465 case aco_opcode::v_madmk_f16: return gfx_level >= GFX9; in instr_is_16bit()
818 if (gfx_level >= GFX9) in wait_imm()
[all …]
Daco_print_asm.cpp135 case GFX9: in to_clrx_device_name()
287 ((gfx_level >= GFX9 && in disasm_instr()
291 (gfx_level <= GFX9 && in disasm_instr()
294 (gfx_level == GFX9 && (binary[pos] & 0xffff8000) == 0xd1ff8000)) /* v_add3_u32 + clamp */) { in disasm_instr()
317 } else if (gfx_level == GFX9 && (binary[pos] & 0xfc024000) == 0xc0024000) { in disasm_instr()
/third_party/mesa3d/src/amd/vulkan/winsys/null/
Dradv_null_winsys.c96 info->gfx_level = GFX9; in radv_null_winsys_query_info()
139 info->has_packed_math_16bit = info->gfx_level >= GFX9; in radv_null_winsys_query_info()
148 info->address32_hi = info->gfx_level >= GFX9 ? 0xffff8000u : 0x0; in radv_null_winsys_query_info()
150 info->has_rbplus = info->family == CHIP_STONEY || info->gfx_level >= GFX9; in radv_null_winsys_query_info()
/third_party/mesa3d/src/amd/vulkan/
Dsi_cmd_buffer.c108 if (device->physical_device->rad_info.gfx_level >= GFX9 && in si_emit_compute()
234 if (physical_device->rad_info.gfx_level <= GFX9) in si_emit_graphics()
298 } else if (physical_device->rad_info.gfx_level == GFX9) { in si_emit_graphics()
322 } else if (device->physical_device->rad_info.gfx_level == GFX9) { in si_emit_graphics()
371 } else if (physical_device->rad_info.gfx_level == GFX9) { in si_emit_graphics()
507 if (physical_device->rad_info.gfx_level >= GFX9) { in si_emit_graphics()
538 if (physical_device->rad_info.gfx_level >= GFX9) { in si_emit_graphics()
935 unsigned is_gfx8_mec = is_mec && gfx_level < GFX9; in si_cs_emit_write_event_eop()
943 if (gfx_level >= GFX9 || is_gfx8_mec) { in si_cs_emit_write_event_eop()
948 if (gfx_level == GFX9 && !is_mec) { in si_cs_emit_write_event_eop()
[all …]
Dradv_image.c89 if (device->physical_device->rad_info.gfx_level < GFX9) { in radv_use_tc_compat_htile_for_image()
113 if (device->physical_device->rad_info.gfx_level >= GFX9) in radv_surface_has_scanout()
279 device->physical_device->rad_info.gfx_level == GFX9) in radv_use_dcc_for_image_early()
426 if (device->physical_device->rad_info.gfx_level >= GFX9) { in radv_patch_surface_from_metadata()
534 if (create_info->prime_blit_src && device->physical_device->rad_info.gfx_level == GFX9) { in radv_patch_image_from_extra_info()
633 if (device->physical_device->rad_info.gfx_level >= GFX9 && in radv_get_surface_flags()
779 if (gfx_level >= GFX9) { in si_set_mutable_tex_desc_fields()
788 if (gfx_level >= GFX9 || base_level_info->mode == RADEON_SURF_MODE_2D) in si_set_mutable_tex_desc_fields()
810 if (gfx_level <= GFX9) in si_set_mutable_tex_desc_fields()
843 } else if (gfx_level == GFX9) { in si_set_mutable_tex_desc_fields()
[all …]
Dradv_wsi.c61 if (device->physical_device->rad_info.gfx_level >= GFX9 && in radv_wsi_get_prime_blit_queue()
99 physical_device->wsi_device.supports_modifiers = physical_device->rad_info.gfx_level >= GFX9; in radv_init_wsi()
Dradv_sqtt.c152 if (device->physical_device->rad_info.gfx_level < GFX9) { in radv_emit_thread_trace_start()
171 if (device->physical_device->rad_info.gfx_level == GFX9) { in radv_emit_thread_trace_start()
183 if (device->physical_device->rad_info.gfx_level == GFX9) { in radv_emit_thread_trace_start()
232 } else if (device->physical_device->rad_info.gfx_level == GFX9) { in radv_copy_thread_trace_info_regs()
370 if (device->physical_device->rad_info.gfx_level >= GFX9) { in radv_emit_spi_config_cntl()
/third_party/mesa3d/src/amd/compiler/tests/
Dtest_assembler.cpp230 for (unsigned i = GFX9; i <= GFX10; i++) {
248 for (unsigned i = GFX9; i <= GFX10; i++) {
267 for (unsigned i = GFX9; i <= GFX10; i++) {
281 if (i >= GFX9) {
292 if (!setup_cs(NULL, GFX9))
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))
556 if (!setup_cs("v1 v1 v2 v1", GFX9))
689 if (!setup_cs("v1 v1 v1", GFX9))
716 for (unsigned i = GFX9; i <= GFX10; i++) {
739 for (unsigned i = GFX8; i <= GFX9; i++) {
868 for (unsigned i = GFX8; i <= GFX9; i++) {
893 …bool can_propagate = cfg.src == aco_opcode::v_rcp_f32 || (i >= GFX9 && cfg.src == aco_opcode::v_mi…
[all …]
Dtest_to_hw_instr.cpp226 for (amd_gfx_level lvl : {GFX8, GFX9, GFX11}) {
421 for (amd_gfx_level lvl : {GFX9, GFX10, GFX11}) {
574 if (!setup_cs(NULL, GFX9))
604 for (amd_gfx_level lvl : {GFX7, GFX8, GFX9, GFX11}) {
703 for (amd_gfx_level lvl : {GFX7, GFX8, GFX9, GFX11}) {
/third_party/mesa3d/src/gallium/drivers/radeonsi/
Dsi_state_draw.cpp40 #define GFX(name) name##GFX9
219 if (GFX_VERSION >= GFX9 && HAS_TESS) in si_update_shaders()
221 else if (GFX_VERSION >= GFX9 && HAS_GS) in si_update_shaders()
277 if ((GFX_VERSION >= GFX10_3 || (GFX_VERSION >= GFX9 && sctx->screen->info.rbplus_allowed)) && in si_update_shaders()
432 if (GFX_VERSION >= GFX9) { in si_cp_dma_prefetch_inline()
464 case GFX9: in si_cp_dma_prefetch()
465 si_cp_dma_prefetch_inline<GFX9>(sctx, buf, offset, size); in si_cp_dma_prefetch()
530 } else if (GFX_VERSION >= GFX9) { in si_prefetch_shaders()
633 if (sctx->gfx_level >= GFX9) { in si_emit_derived_tess_state()
806 if (sctx->gfx_level >= GFX9) { in si_emit_derived_tess_state()
[all …]
Dsi_build_pm4.h159 if ((gfx_level) < GFX9 || \
160 ((gfx_level) == GFX9 && (screen)->info.me_fw_version < 26)) \
336 } else if (gfx_level == GFX9) { in si_get_user_data_base()
354 if (gfx_level == GFX9) { in si_get_user_data_base()
379 if (gfx_level == GFX9) { in si_get_user_data_base()
Dsi_fence.c80 if (ctx->gfx_level >= GFX9 || (compute_ib && ctx->gfx_level >= GFX7)) { in si_cp_release_mem()
88 if (ctx->gfx_level == GFX9 && !compute_ib && query_type != PIPE_QUERY_OCCLUSION_COUNTER && in si_cp_release_mem()
120 radeon_emit(PKT3(PKT3_RELEASE_MEM, ctx->gfx_level >= GFX9 ? 6 : 5, 0)); in si_cp_release_mem()
127 if (ctx->gfx_level >= GFX9) in si_cp_release_mem()
/third_party/mesa3d/src/intel/compiler/
Dbrw_gfx_ver_enum.h38 GFX9 = (1 << 7), enumerator
61 case 90: return GFX9; in gfx_ver_from_devinfo()
/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

1234567