[Mesa-dev] [PATCH 2/2] ac: rename SI-CIK-VI to GFX6-GFX7-GFX8
Marek Olšák
maraeo at gmail.com
Wed May 15 02:17:34 UTC 2019
From: Marek Olšák <marek.olsak at amd.com>
---
src/amd/common/ac_debug.c | 2 +-
src/amd/common/ac_gpu_info.c | 24 ++--
src/amd/common/ac_gpu_info.h | 2 +-
src/amd/common/ac_llvm_build.c | 26 ++---
src/amd/common/ac_nir_to_llvm.c | 34 +++---
src/amd/common/ac_shader_abi.h | 2 +-
src/amd/common/ac_shader_util.c | 6 +-
src/amd/common/ac_surface.c | 16 +--
src/amd/common/ac_surface.h | 2 +-
src/amd/common/amd_family.h | 6 +-
src/amd/vulkan/radv_cmd_buffer.c | 34 +++---
src/amd/vulkan/radv_debug.c | 2 +-
src/amd/vulkan/radv_device.c | 84 +++++++-------
src/amd/vulkan/radv_extensions.py | 12 +-
src/amd/vulkan/radv_formats.c | 2 +-
src/amd/vulkan/radv_image.c | 18 +--
src/amd/vulkan/radv_nir_to_llvm.c | 16 +--
src/amd/vulkan/radv_pipeline.c | 34 +++---
src/amd/vulkan/radv_private.h | 2 +-
src/amd/vulkan/radv_shader.c | 4 +-
src/amd/vulkan/si_cmd_buffer.c | 66 +++++------
src/amd/vulkan/winsys/amdgpu/radv_amdgpu_cs.c | 2 +-
.../vulkan/winsys/amdgpu/radv_amdgpu_winsys.c | 2 +-
src/gallium/drivers/r600/r600_texture.c | 4 +-
src/gallium/drivers/r600/r600d_common.h | 2 +-
src/gallium/drivers/radeonsi/cik_sdma.c | 26 ++---
src/gallium/drivers/radeonsi/si_blit.c | 2 +-
src/gallium/drivers/radeonsi/si_clear.c | 8 +-
src/gallium/drivers/radeonsi/si_compute.c | 26 ++---
.../drivers/radeonsi/si_compute_blit.c | 8 +-
src/gallium/drivers/radeonsi/si_cp_dma.c | 14 +--
src/gallium/drivers/radeonsi/si_debug.c | 2 +-
src/gallium/drivers/radeonsi/si_descriptors.c | 14 +--
src/gallium/drivers/radeonsi/si_dma_cs.c | 8 +-
src/gallium/drivers/radeonsi/si_fence.c | 8 +-
src/gallium/drivers/radeonsi/si_get.c | 2 +-
src/gallium/drivers/radeonsi/si_gfx_cs.c | 6 +-
src/gallium/drivers/radeonsi/si_gpu_load.c | 4 +-
src/gallium/drivers/radeonsi/si_perfcounter.c | 8 +-
src/gallium/drivers/radeonsi/si_pipe.c | 46 ++++----
src/gallium/drivers/radeonsi/si_pipe.h | 10 +-
src/gallium/drivers/radeonsi/si_pm4.c | 4 +-
src/gallium/drivers/radeonsi/si_query.c | 10 +-
src/gallium/drivers/radeonsi/si_shader.c | 28 ++---
src/gallium/drivers/radeonsi/si_shader.h | 2 +-
.../drivers/radeonsi/si_shader_tgsi_mem.c | 24 ++--
src/gallium/drivers/radeonsi/si_state.c | 108 +++++++++---------
src/gallium/drivers/radeonsi/si_state_draw.c | 86 +++++++-------
.../drivers/radeonsi/si_state_shaders.c | 48 ++++----
.../drivers/radeonsi/si_state_streamout.c | 6 +-
.../drivers/radeonsi/si_state_viewport.c | 8 +-
.../drivers/radeonsi/si_test_dma_perf.c | 12 +-
src/gallium/drivers/radeonsi/si_texture.c | 36 +++---
src/gallium/winsys/amdgpu/drm/amdgpu_cs.c | 4 +-
src/gallium/winsys/amdgpu/drm/amdgpu_winsys.c | 6 +-
src/gallium/winsys/radeon/drm/radeon_drm_cs.c | 2 +-
.../winsys/radeon/drm/radeon_drm_surface.c | 6 +-
.../winsys/radeon/drm/radeon_drm_winsys.c | 30 ++---
src/mesa/state_tracker/st_draw.c | 2 +-
59 files changed, 509 insertions(+), 509 deletions(-)
diff --git a/src/amd/common/ac_debug.c b/src/amd/common/ac_debug.c
index e5463b66616..187e9d6ba66 100644
--- a/src/amd/common/ac_debug.c
+++ b/src/amd/common/ac_debug.c
@@ -269,5 +269,5 @@ static void ac_parse_packet3(FILE *f, uint32_t header, struct ac_ib_parser *ib,
break;
case PKT3_SURFACE_SYNC:
- if (ib->chip_class >= CIK) {
+ if (ib->chip_class >= GFX7) {
ac_dump_reg(f, ib->chip_class, R_0301F0_CP_COHER_CNTL, ac_ib_get(ib), ~0);
ac_dump_reg(f, ib->chip_class, R_0301F4_CP_COHER_SIZE, ac_ib_get(ib), ~0);
diff --git a/src/amd/common/ac_gpu_info.c b/src/amd/common/ac_gpu_info.c
index 02c2086f684..171560d6f04 100644
--- a/src/amd/common/ac_gpu_info.c
+++ b/src/amd/common/ac_gpu_info.c
@@ -79,5 +79,5 @@ static unsigned cik_get_num_tile_pipes(struct amdgpu_gpu_info *info)
return 16;
default:
- fprintf(stderr, "Invalid CIK pipe configuration, assuming P2\n");
+ fprintf(stderr, "Invalid GFX7 pipe configuration, assuming P2\n");
assert(!"this should never occur");
return 2;
@@ -324,9 +324,9 @@ bool ac_query_gpu_info(int fd, amdgpu_device_handle dev,
info->chip_class = GFX9;
else if (info->family >= CHIP_TONGA)
- info->chip_class = VI;
+ info->chip_class = GFX8;
else if (info->family >= CHIP_BONAIRE)
- info->chip_class = CIK;
+ info->chip_class = GFX7;
else if (info->family >= CHIP_TAHITI)
- info->chip_class = SI;
+ info->chip_class = GFX6;
else {
fprintf(stderr, "amdgpu: Unknown family.\n");
@@ -382,16 +382,16 @@ bool ac_query_gpu_info(int fd, amdgpu_device_handle dev,
info->has_eqaa_surface_allocator = true;
info->has_format_bc1_through_bc7 = true;
- /* DRM 3.1.0 doesn't flush TC for VI correctly. */
- info->kernel_flushes_tc_l2_after_ib = info->chip_class != VI ||
+ /* DRM 3.1.0 doesn't flush TC for GFX8 correctly. */
+ info->kernel_flushes_tc_l2_after_ib = info->chip_class != GFX8 ||
info->drm_minor >= 2;
info->has_indirect_compute_dispatch = true;
- /* SI doesn't support unaligned loads. */
- info->has_unaligned_shader_loads = info->chip_class != SI;
- /* Disable sparse mappings on SI due to VM faults in CP DMA. Enable them once
+ /* GFX6 doesn't support unaligned loads. */
+ info->has_unaligned_shader_loads = info->chip_class != GFX6;
+ /* Disable sparse mappings on GFX6 due to VM faults in CP DMA. Enable them once
* these faults are mitigated in software.
* Disable sparse mappings on GFX9 due to hangs.
*/
info->has_sparse_vm_mappings =
- info->chip_class >= CIK && info->chip_class <= VI &&
+ info->chip_class >= GFX7 && info->chip_class <= GFX8 &&
info->drm_minor >= 13;
info->has_2d_tiling = true;
@@ -446,5 +446,5 @@ bool ac_query_gpu_info(int fd, amdgpu_device_handle dev,
info->gart_page_size = alignment_info.size_remote;
- if (info->chip_class == SI)
+ if (info->chip_class == GFX6)
info->gfx_ib_pad_with_type2 = TRUE;
@@ -790,5 +790,5 @@ ac_get_harvested_configs(struct radeon_info *info,
- if (info->chip_class >= CIK) {
+ if (info->chip_class >= GFX7) {
unsigned raster_config_1 = *cik_raster_config_1_p;
if ((num_se > 2) && ((!se_mask[0] && !se_mask[1]) ||
diff --git a/src/amd/common/ac_gpu_info.h b/src/amd/common/ac_gpu_info.h
index 25bd33c1afa..82f8f1e8c75 100644
--- a/src/amd/common/ac_gpu_info.h
+++ b/src/amd/common/ac_gpu_info.h
@@ -183,5 +183,5 @@ static inline uint32_t
ac_get_num_physical_sgprs(enum chip_class chip_class)
{
- return chip_class >= VI ? 800 : 512;
+ return chip_class >= GFX8 ? 800 : 512;
}
diff --git a/src/amd/common/ac_llvm_build.c b/src/amd/common/ac_llvm_build.c
index 58dcd2e863d..3ad9bb34805 100644
--- a/src/amd/common/ac_llvm_build.c
+++ b/src/amd/common/ac_llvm_build.c
@@ -827,5 +827,5 @@ ac_prepare_cube_coords(struct ac_llvm_context *ctx,
* layer due to extrapolation."
*
- * VI and earlier attempt to implement this in hardware by
+ * GFX8 and earlier attempt to implement this in hardware by
* clamping the value of coords[2] = (8 * layer) + face.
* Unfortunately, this means that the we end up with the wrong
@@ -834,5 +834,5 @@ ac_prepare_cube_coords(struct ac_llvm_context *ctx,
* Clamp the layer earlier to work around the issue.
*/
- if (ctx->chip_class <= VI) {
+ if (ctx->chip_class <= GFX8) {
LLVMValueRef ge0;
ge0 = LLVMBuildFCmp(builder, LLVMRealOGE, tmp, ctx->f32_0, "");
@@ -1393,5 +1393,5 @@ ac_build_buffer_load(struct ac_llvm_context *ctx,
if (allow_smem && !slc &&
- (!glc || (HAVE_LLVM >= 0x0800 && ctx->chip_class >= VI))) {
+ (!glc || (HAVE_LLVM >= 0x0800 && ctx->chip_class >= GFX8))) {
assert(vindex == NULL);
@@ -1784,5 +1784,5 @@ ac_build_opencoded_load_format(struct ac_llvm_context *ctx,
int log_recombine = 0;
- if (ctx->chip_class == SI && !known_aligned) {
+ if (ctx->chip_class == GFX6 && !known_aligned) {
/* Avoid alignment restrictions by loading one byte at a time. */
load_num_channels <<= load_log_size;
@@ -1820,5 +1820,5 @@ ac_build_opencoded_load_format(struct ac_llvm_context *ctx,
if (log_recombine > 0) {
- /* Recombine bytes if necessary (SI only) */
+ /* Recombine bytes if necessary (GFX6 only) */
LLVMTypeRef dst_type = log_recombine == 2 ? ctx->i32 : ctx->i16;
@@ -2213,5 +2213,5 @@ ac_get_thread_id(struct ac_llvm_context *ctx)
/*
- * SI implements derivatives using the local data store (LDS)
+ * AMD GCN implements derivatives using the local data store (LDS)
* All writes to the LDS happen in all executing threads at
* the same time. TID is the Thread ID for the current
@@ -3305,5 +3305,5 @@ void ac_init_exec_full_mask(struct ac_llvm_context *ctx)
void ac_declare_lds_as_pointer(struct ac_llvm_context *ctx)
{
- unsigned lds_size = ctx->chip_class >= CIK ? 65536 : 32768;
+ unsigned lds_size = ctx->chip_class >= GFX7 ? 65536 : 32768;
ctx->lds = LLVMBuildIntToPtr(ctx->builder, ctx->i32_0,
LLVMPointerType(LLVMArrayType(ctx->i32, lds_size / 4), AC_ADDR_SPACE_LDS),
@@ -4035,5 +4035,5 @@ ac_build_alu_op(struct ac_llvm_context *ctx, LLVMValueRef lhs, LLVMValueRef rhs,
* prefix of this many threads
*
- * TODO: add inclusive and excluse scan functions for SI chip class.
+ * TODO: add inclusive and excluse scan functions for GFX6.
*/
static LLVMValueRef
@@ -4143,5 +4143,5 @@ ac_build_reduce(struct ac_llvm_context *ctx, LLVMValueRef src, nir_op op, unsign
if (cluster_size == 4) return ac_build_wwm(ctx, result);
- if (ctx->chip_class >= VI)
+ if (ctx->chip_class >= GFX8)
swap = ac_build_dpp(ctx, identity, result, dpp_row_half_mirror, 0xf, 0xf, false);
else
@@ -4150,5 +4150,5 @@ ac_build_reduce(struct ac_llvm_context *ctx, LLVMValueRef src, nir_op op, unsign
if (cluster_size == 8) return ac_build_wwm(ctx, result);
- if (ctx->chip_class >= VI)
+ if (ctx->chip_class >= GFX8)
swap = ac_build_dpp(ctx, identity, result, dpp_row_mirror, 0xf, 0xf, false);
else
@@ -4157,5 +4157,5 @@ ac_build_reduce(struct ac_llvm_context *ctx, LLVMValueRef src, nir_op op, unsign
if (cluster_size == 16) return ac_build_wwm(ctx, result);
- if (ctx->chip_class >= VI && cluster_size != 32)
+ if (ctx->chip_class >= GFX8 && cluster_size != 32)
swap = ac_build_dpp(ctx, identity, result, dpp_row_bcast15, 0xa, 0xf, false);
else
@@ -4164,5 +4164,5 @@ ac_build_reduce(struct ac_llvm_context *ctx, LLVMValueRef src, nir_op op, unsign
if (cluster_size == 32) return ac_build_wwm(ctx, result);
- if (ctx->chip_class >= VI) {
+ if (ctx->chip_class >= GFX8) {
swap = ac_build_dpp(ctx, identity, result, dpp_row_bcast31, 0xc, 0xf, false);
result = ac_build_alu_op(ctx, result, swap, op);
@@ -4351,5 +4351,5 @@ ac_build_quad_swizzle(struct ac_llvm_context *ctx, LLVMValueRef src,
{
unsigned mask = dpp_quad_perm(lane0, lane1, lane2, lane3);
- if (ctx->chip_class >= VI) {
+ if (ctx->chip_class >= GFX8) {
return ac_build_dpp(ctx, src, src, mask, 0xf, 0xf, false);
} else {
diff --git a/src/amd/common/ac_nir_to_llvm.c b/src/amd/common/ac_nir_to_llvm.c
index 682645e9b1f..53c4ff7d383 100644
--- a/src/amd/common/ac_nir_to_llvm.c
+++ b/src/amd/common/ac_nir_to_llvm.c
@@ -113,5 +113,5 @@ get_ac_image_dim(const struct ac_llvm_context *ctx, enum glsl_sampler_dim sdim,
if (dim == ac_image_cube ||
- (ctx->chip_class <= VI && dim == ac_image_3d))
+ (ctx->chip_class <= GFX8 && dim == ac_image_3d))
dim = ac_image_2darray;
@@ -372,5 +372,5 @@ static LLVMValueRef emit_f2f16(struct ac_llvm_context *ctx,
result = LLVMBuildFPTrunc(ctx->builder, src0, ctx->f16, "");
- if (ctx->chip_class >= VI) {
+ if (ctx->chip_class >= GFX8) {
LLVMValueRef args[2];
/* Check if the result is a denormal - and flush to 0 if so. */
@@ -383,8 +383,8 @@ static LLVMValueRef emit_f2f16(struct ac_llvm_context *ctx,
result = LLVMBuildFPExt(ctx->builder, result, ctx->f32, "");
- if (ctx->chip_class >= VI)
+ if (ctx->chip_class >= GFX8)
result = LLVMBuildSelect(ctx->builder, cond, ctx->f32_0, result, "");
else {
- /* for SI/CIK */
+ /* for GFX6-GFX7 */
/* 0x38800000 is smallest half float value (2^-14) in 32-bit float,
* so compare the result and flush to 0 if it's smaller.
@@ -1170,7 +1170,7 @@ get_buffer_size(struct ac_nir_context *ctx, LLVMValueRef descriptor, bool in_ele
LLVMConstInt(ctx->ac.i32, 2, false), "");
- /* VI only */
- if (ctx->ac.chip_class == VI && in_elements) {
- /* On VI, the descriptor contains the size in bytes,
+ /* GFX8 only */
+ if (ctx->ac.chip_class == GFX8 && in_elements) {
+ /* On GFX8, the descriptor contains the size in bytes,
* but TXQ must return the size in elements.
* The stride is always non-zero for resources using TXQ.
@@ -1377,5 +1377,5 @@ static LLVMValueRef build_tex_intrinsic(struct ac_nir_context *ctx,
}
- if (instr->op == nir_texop_tg4 && ctx->ac.chip_class <= VI) {
+ if (instr->op == nir_texop_tg4 && ctx->ac.chip_class <= GFX8) {
nir_deref_instr *texture_deref_instr = get_tex_texture_deref(instr);
nir_variable *var = nir_deref_instr_get_variable(texture_deref_instr);
@@ -1536,9 +1536,9 @@ static unsigned get_cache_policy(struct ac_nir_context *ctx,
unsigned cache_policy = 0;
- /* SI has a TC L1 bug causing corruption of 8bit/16bit stores. All
+ /* GFX6 has a TC L1 bug causing corruption of 8bit/16bit stores. All
* store opcodes not aligned to a dword are affected. The only way to
* get unaligned stores is through shader images.
*/
- if (((may_store_unaligned && ctx->ac.chip_class == SI) ||
+ if (((may_store_unaligned && ctx->ac.chip_class == GFX6) ||
/* If this is write-only, don't keep data in L1 to prevent
* evicting L1 cache lines that may be needed by other
@@ -2774,9 +2774,9 @@ static void emit_membar(struct ac_llvm_context *ac,
void ac_emit_barrier(struct ac_llvm_context *ac, gl_shader_stage stage)
{
- /* SI only (thanks to a hw bug workaround):
+ /* GFX6 only (thanks to a hw bug workaround):
* The real barrier instruction isn’t needed, because an entire patch
* always fits into a single wave.
*/
- if (ac->chip_class == SI && stage == MESA_SHADER_TESS_CTRL) {
+ if (ac->chip_class == GFX6 && stage == MESA_SHADER_TESS_CTRL) {
ac_build_waitcnt(ac, LGKM_CNT & VM_CNT);
return;
@@ -3558,5 +3558,5 @@ static LLVMValueRef get_sampler_desc(struct ac_nir_context *ctx,
/* Disable anisotropic filtering if BASE_LEVEL == LAST_LEVEL.
*
- * SI-CI:
+ * GFX6-GFX7:
* If BASE_LEVEL == LAST_LEVEL, the shader must disable anisotropic
* filtering manually. The driver sets img7 to a mask clearing
@@ -3564,5 +3564,5 @@ static LLVMValueRef get_sampler_desc(struct ac_nir_context *ctx,
* s_and_b32 samp0, samp0, img7
*
- * VI:
+ * GFX8:
* The ANISO_OVERRIDE sampler field enables this fix in TA.
*/
@@ -3573,5 +3573,5 @@ static LLVMValueRef sici_fix_sampler_aniso(struct ac_nir_context *ctx,
LLVMValueRef img7, samp0;
- if (ctx->ac.chip_class >= VI)
+ if (ctx->ac.chip_class >= GFX8)
return samp;
@@ -3757,5 +3757,5 @@ static void visit_tex(struct ac_nir_context *ctx, nir_tex_instr *instr)
* Z32_FLOAT, but we don't know that here.
*/
- if (args.compare && ctx->ac.chip_class >= VI && ctx->abi->clamp_shadow_reference)
+ if (args.compare && ctx->ac.chip_class >= GFX8 && ctx->abi->clamp_shadow_reference)
args.compare = ac_build_clamp(&ctx->ac, ac_to_float(&ctx->ac, args.compare));
@@ -4397,5 +4397,5 @@ ac_lower_indirect_derefs(struct nir_shader *nir, enum chip_class chip_class)
* on GFX9.
*/
- bool llvm_has_working_vgpr_indexing = chip_class <= VI;
+ bool llvm_has_working_vgpr_indexing = chip_class <= GFX8;
/* TODO: Indirect indexing of GS inputs is unimplemented.
diff --git a/src/amd/common/ac_shader_abi.h b/src/amd/common/ac_shader_abi.h
index 8debb1ff986..2051f22d29b 100644
--- a/src/amd/common/ac_shader_abi.h
+++ b/src/amd/common/ac_shader_abi.h
@@ -197,5 +197,5 @@ struct ac_shader_abi {
LLVMValueRef (*load_base_vertex)(struct ac_shader_abi *abi);
- /* Whether to clamp the shadow reference value to [0,1]on VI. Radeonsi currently
+ /* Whether to clamp the shadow reference value to [0,1]on GFX8. Radeonsi currently
* uses it due to promoting D16 to D32, but radv needs it off. */
bool clamp_shadow_reference;
diff --git a/src/amd/common/ac_shader_util.c b/src/amd/common/ac_shader_util.c
index 531395f4f62..64152081737 100644
--- a/src/amd/common/ac_shader_util.c
+++ b/src/amd/common/ac_shader_util.c
@@ -105,5 +105,5 @@ ac_vgt_gs_mode(unsigned gs_max_vert_out, enum chip_class chip_class)
return S_028A40_MODE(V_028A40_GS_SCENARIO_G) |
S_028A40_CUT_MODE(cut_mode)|
- S_028A40_ES_WRITE_OPTIMIZE(chip_class <= VI) |
+ S_028A40_ES_WRITE_OPTIMIZE(chip_class <= GFX8) |
S_028A40_GS_WRITE_OPTIMIZE(1) |
S_028A40_ONCHIP(chip_class >= GFX9 ? 1 : 0);
@@ -168,7 +168,7 @@ ac_export_mrt_z(struct ac_llvm_context *ctx, LLVMValueRef depth,
}
- /* SI (except OLAND and HAINAN) has a bug that it only looks
+ /* GFX6 (except OLAND and HAINAN) has a bug that it only looks
* at the X writemask component. */
- if (ctx->chip_class == SI &&
+ if (ctx->chip_class == GFX6 &&
ctx->family != CHIP_OLAND &&
ctx->family != CHIP_HAINAN)
diff --git a/src/amd/common/ac_surface.c b/src/amd/common/ac_surface.c
index f9dd4f5d77d..a9433b9696c 100644
--- a/src/amd/common/ac_surface.c
+++ b/src/amd/common/ac_surface.c
@@ -453,5 +453,5 @@ static void gfx6_set_micro_tile_mode(struct radeon_surf *surf,
uint32_t tile_mode = info->si_tile_mode_array[surf->u.legacy.tiling_index[0]];
- if (info->chip_class >= CIK)
+ if (info->chip_class >= GFX7)
surf->micro_tile_mode = G_009910_MICRO_TILE_MODE_NEW(tile_mode);
else
@@ -527,6 +527,6 @@ static int gfx6_surface_settings(ADDR_HANDLE addrlib,
/* Compute tile swizzle. */
- /* TODO: fix tile swizzle with mipmapping for SI */
- if ((info->chip_class >= CIK || config->info.levels == 1) &&
+ /* TODO: fix tile swizzle with mipmapping for GFX6 */
+ if ((info->chip_class >= GFX7 || config->info.levels == 1) &&
config->info.surf_index &&
surf->u.legacy.level[0].mode == RADEON_SURF_MODE_2D &&
@@ -568,5 +568,5 @@ void ac_compute_cmask(const struct radeon_info *info,
return;
- assert(info->chip_class <= VI);
+ assert(info->chip_class <= GFX8);
switch (num_pipes) {
@@ -733,5 +733,5 @@ static int gfx6_compute_surface(ADDR_HANDLE addrlib,
*/
AddrSurfInfoIn.flags.dccCompatible =
- info->chip_class >= VI &&
+ info->chip_class >= GFX8 &&
!(surf->flags & RADEON_SURF_Z_OR_SBUFFER) &&
!(surf->flags & RADEON_SURF_DISABLE_DCC) &&
@@ -743,5 +743,5 @@ static int gfx6_compute_surface(ADDR_HANDLE addrlib,
AddrSurfInfoIn.flags.compressZ = !!(surf->flags & RADEON_SURF_Z_OR_SBUFFER);
- /* On CI/VI, the DB uses the same pitch and tile mode (except tilesplit)
+ /* On GFX7-GFX8, the DB uses the same pitch and tile mode (except tilesplit)
* for Z and stencil. This can cause a number of problems which we work
* around here:
@@ -800,5 +800,5 @@ static int gfx6_compute_surface(ADDR_HANDLE addrlib,
assert(AddrSurfInfoIn.tileMode == ADDR_TM_2D_TILED_THIN1);
- if (info->chip_class == SI) {
+ if (info->chip_class == GFX6) {
if (AddrSurfInfoIn.tileType == ADDR_DISPLAYABLE) {
if (surf->bpe == 2)
@@ -817,5 +817,5 @@ static int gfx6_compute_surface(ADDR_HANDLE addrlib,
}
} else {
- /* CIK - VI */
+ /* GFX7 - GFX8 */
if (AddrSurfInfoIn.tileType == ADDR_DISPLAYABLE)
AddrSurfInfoIn.tileIndex = 10; /* 2D displayable */
diff --git a/src/amd/common/ac_surface.h b/src/amd/common/ac_surface.h
index 10d25e23d32..08aac94d3a9 100644
--- a/src/amd/common/ac_surface.h
+++ b/src/amd/common/ac_surface.h
@@ -222,5 +222,5 @@ struct radeon_surf {
union {
- /* R600-VI return values.
+ /* Return values for GFX8 and older.
*
* Some of them can be set by the caller if certain parameters are
diff --git a/src/amd/common/amd_family.h b/src/amd/common/amd_family.h
index eed6553d44b..1a95f93b608 100644
--- a/src/amd/common/amd_family.h
+++ b/src/amd/common/amd_family.h
@@ -111,7 +111,7 @@ enum chip_class {
EVERGREEN,
CAYMAN,
- SI, /* GFX6 */
- CIK, /* GFX7 */
- VI, /* GFX8 */
+ GFX6,
+ GFX7,
+ GFX8,
GFX9,
};
diff --git a/src/amd/vulkan/radv_cmd_buffer.c b/src/amd/vulkan/radv_cmd_buffer.c
index 6d1f3fc7d5a..ec1fcf4fd64 100644
--- a/src/amd/vulkan/radv_cmd_buffer.c
+++ b/src/amd/vulkan/radv_cmd_buffer.c
@@ -216,5 +216,5 @@ bool radv_cmd_buffer_uses_mec(struct radv_cmd_buffer *cmd_buffer)
{
return cmd_buffer->queue_family_index == RADV_QUEUE_COMPUTE &&
- cmd_buffer->device->physical_device->rad_info.chip_class >= CIK;
+ cmd_buffer->device->physical_device->rad_info.chip_class >= GFX7;
}
@@ -1042,5 +1042,5 @@ radv_emit_fb_color_state(struct radv_cmd_buffer *cmd_buffer,
VkImageLayout layout)
{
- bool is_vi = cmd_buffer->device->physical_device->rad_info.chip_class >= VI;
+ bool is_vi = cmd_buffer->device->physical_device->rad_info.chip_class >= GFX8;
struct radv_color_buffer_info *cb = &att->cb;
uint32_t cb_color_info = cb->cb_color_info;
@@ -1630,6 +1630,6 @@ radv_emit_framebuffer_state(struct radv_cmd_buffer *cmd_buffer)
S_028208_BR_Y(framebuffer->height));
- if (cmd_buffer->device->physical_device->rad_info.chip_class >= VI) {
- uint8_t watermark = 4; /* Default value for VI. */
+ if (cmd_buffer->device->physical_device->rad_info.chip_class >= GFX8) {
+ uint8_t watermark = 4; /* Default value for GFX8. */
/* For optimal DCC performance. */
@@ -1692,5 +1692,5 @@ void radv_set_db_count_control(struct radv_cmd_buffer *cmd_buffer)
if(!cmd_buffer->state.active_occlusion_queries) {
- if (cmd_buffer->device->physical_device->rad_info.chip_class >= CIK) {
+ if (cmd_buffer->device->physical_device->rad_info.chip_class >= GFX7) {
if (G_028A4C_OUT_OF_ORDER_PRIMITIVE_ENABLE(pa_sc_mode_cntl_1) &&
pipeline->graphics.disable_out_of_order_rast_for_occlusion &&
@@ -1711,5 +1711,5 @@ void radv_set_db_count_control(struct radv_cmd_buffer *cmd_buffer)
uint32_t sample_rate = subpass ? util_logbase2(subpass->max_sample_count) : 0;
- if (cmd_buffer->device->physical_device->rad_info.chip_class >= CIK) {
+ if (cmd_buffer->device->physical_device->rad_info.chip_class >= GFX7) {
db_count_control =
S_028004_PERFECT_ZPASS_COUNTS(has_perfect_queries) |
@@ -2020,5 +2020,5 @@ radv_flush_vertex_descriptors(struct radv_cmd_buffer *cmd_buffer,
desc[0] = va;
desc[1] = S_008F04_BASE_ADDRESS_HI(va >> 32) | S_008F04_STRIDE(stride);
- if (cmd_buffer->device->physical_device->rad_info.chip_class <= CIK && stride)
+ if (cmd_buffer->device->physical_device->rad_info.chip_class <= GFX7 && stride)
desc[2] = (buffer->size - offset - velems->format_size[i]) / stride + 1;
else
@@ -2107,5 +2107,5 @@ radv_flush_streamout_descriptors(struct radv_cmd_buffer *cmd_buffer)
/* Set the descriptor.
*
- * On VI, the format must be non-INVALID, otherwise
+ * On GFX8, the format must be non-INVALID, otherwise
* the buffer will be considered not bound and store
* instructions will be no-ops.
@@ -2212,5 +2212,5 @@ radv_emit_draw_registers(struct radv_cmd_buffer *cmd_buffer,
R_030960_IA_MULTI_VGT_PARAM,
4, ia_multi_vgt_param);
- } else if (info->chip_class >= CIK) {
+ } else if (info->chip_class >= GFX7) {
radeon_set_context_reg_idx(cs,
R_028AA8_IA_MULTI_VGT_PARAM,
@@ -2949,5 +2949,5 @@ VkResult radv_EndCommandBuffer(
if (cmd_buffer->queue_family_index != RADV_QUEUE_TRANSFER) {
- if (cmd_buffer->device->physical_device->rad_info.chip_class == SI)
+ if (cmd_buffer->device->physical_device->rad_info.chip_class == GFX6)
cmd_buffer->state.flush_bits |= RADV_CMD_FLAG_CS_PARTIAL_FLUSH | RADV_CMD_FLAG_PS_PARTIAL_FLUSH | RADV_CMD_FLAG_WRITEBACK_GLOBAL_L2;
si_emit_cache_flush(cmd_buffer);
@@ -3825,9 +3825,9 @@ radv_emit_all_graphics_states(struct radv_cmd_buffer *cmd_buffer,
radv_emit_index_buffer(cmd_buffer);
} else {
- /* On CI and later, non-indexed draws overwrite VGT_INDEX_TYPE,
+ /* On GFX7 and later, non-indexed draws overwrite VGT_INDEX_TYPE,
* so the state must be re-emitted before the next indexed
* draw.
*/
- if (cmd_buffer->device->physical_device->rad_info.chip_class >= CIK) {
+ if (cmd_buffer->device->physical_device->rad_info.chip_class >= GFX7) {
cmd_buffer->state.last_index_type = -1;
cmd_buffer->state.dirty |= RADV_CMD_DIRTY_INDEX_BUFFER;
@@ -3850,5 +3850,5 @@ radv_draw(struct radv_cmd_buffer *cmd_buffer,
&cmd_buffer->device->physical_device->rad_info;
bool has_prefetch =
- cmd_buffer->device->physical_device->rad_info.chip_class >= CIK;
+ cmd_buffer->device->physical_device->rad_info.chip_class >= GFX7;
bool pipeline_is_dirty =
(cmd_buffer->state.dirty & RADV_CMD_DIRTY_PIPELINE) &&
@@ -3860,5 +3860,5 @@ radv_draw(struct radv_cmd_buffer *cmd_buffer,
if (likely(!info->indirect)) {
- /* SI-CI treat instance_count==0 as instance_count==1. There is
+ /* GFX6-GFX7 treat instance_count==0 as instance_count==1. There is
* no workaround for indirect draws, but we can at least skip
* direct draws.
@@ -4238,5 +4238,5 @@ radv_dispatch(struct radv_cmd_buffer *cmd_buffer,
struct radv_pipeline *pipeline = cmd_buffer->state.compute_pipeline;
bool has_prefetch =
- cmd_buffer->device->physical_device->rad_info.chip_class >= CIK;
+ cmd_buffer->device->physical_device->rad_info.chip_class >= GFX7;
bool pipeline_is_dirty = pipeline &&
pipeline != cmd_buffer->state.emitted_compute_pipeline;
@@ -5044,5 +5044,5 @@ static void radv_flush_vgt_streamout(struct radv_cmd_buffer *cmd_buffer)
/* The register is at different places on different ASICs. */
- if (cmd_buffer->device->physical_device->rad_info.chip_class >= CIK) {
+ if (cmd_buffer->device->physical_device->rad_info.chip_class >= GFX7) {
reg_strmout_cntl = R_0300FC_CP_STRMOUT_CNTL;
radeon_set_uconfig_reg(cs, reg_strmout_cntl, 0);
@@ -5085,5 +5085,5 @@ void radv_CmdBeginTransformFeedbackEXT(
counter_buffer_idx = -1;
- /* SI binds streamout buffers as shader resources.
+ /* AMD GCN binds streamout buffers as shader resources.
* VGT only counts primitives and tells the shader through
* SGPRs what to do.
diff --git a/src/amd/vulkan/radv_debug.c b/src/amd/vulkan/radv_debug.c
index 4854b094ba7..432e65b1475 100644
--- a/src/amd/vulkan/radv_debug.c
+++ b/src/amd/vulkan/radv_debug.c
@@ -132,5 +132,5 @@ radv_dump_debug_registers(struct radv_device *device, FILE *f)
radv_dump_mmapped_reg(device, f, R_00D034_SDMA0_STATUS_REG);
radv_dump_mmapped_reg(device, f, R_00D834_SDMA1_STATUS_REG);
- if (info->chip_class <= VI) {
+ if (info->chip_class <= GFX8) {
radv_dump_mmapped_reg(device, f, R_000E50_SRBM_STATUS);
radv_dump_mmapped_reg(device, f, R_000E4C_SRBM_STATUS2);
diff --git a/src/amd/vulkan/radv_device.c b/src/amd/vulkan/radv_device.c
index 4b64f5101ed..c0e317a97e5 100644
--- a/src/amd/vulkan/radv_device.c
+++ b/src/amd/vulkan/radv_device.c
@@ -221,9 +221,9 @@ radv_handle_env_var_force_family(struct radv_physical_device *device)
device->rad_info.chip_class = GFX9;
else if (i >= CHIP_TONGA)
- device->rad_info.chip_class = VI;
+ device->rad_info.chip_class = GFX8;
else if (i >= CHIP_BONAIRE)
- device->rad_info.chip_class = CIK;
+ device->rad_info.chip_class = GFX7;
else
- device->rad_info.chip_class = SI;
+ device->rad_info.chip_class = GFX6;
return;
@@ -333,5 +333,5 @@ radv_physical_device_init(struct radv_physical_device *device,
device->disk_cache = disk_cache_create(device->name, buf, shader_env_flags);
- if (device->rad_info.chip_class < VI ||
+ if (device->rad_info.chip_class < GFX8 ||
device->rad_info.chip_class > GFX9)
fprintf(stderr, "WARNING: radv is not a conformant vulkan implementation, testing use only.\n");
@@ -350,9 +350,9 @@ radv_physical_device_init(struct radv_physical_device *device,
/* The mere presence of CLEAR_STATE in the IB causes random GPU hangs
- * on SI.
+ * on GFX6.
*/
- device->has_clear_state = device->rad_info.chip_class >= CIK;
+ device->has_clear_state = device->rad_info.chip_class >= GFX7;
- device->cpdma_prefetch_writes_memory = device->rad_info.chip_class <= VI;
+ device->cpdma_prefetch_writes_memory = device->rad_info.chip_class <= GFX8;
/* Vega10/Raven need a special workaround for a hardware bug. */
@@ -361,5 +361,5 @@ radv_physical_device_init(struct radv_physical_device *device,
/* Out-of-order primitive rasterization. */
- device->has_out_of_order_rast = device->rad_info.chip_class >= VI &&
+ device->has_out_of_order_rast = device->rad_info.chip_class >= GFX8 &&
device->rad_info.max_se >= 2;
device->out_of_order_rast_allowed = device->has_out_of_order_rast &&
@@ -369,7 +369,7 @@ radv_physical_device_init(struct radv_physical_device *device,
(device->instance->perftest_flags & RADV_PERFTEST_DCC_MSAA);
- /* TODO: Figure out how to use LOAD_CONTEXT_REG on SI/CIK. */
+ /* TODO: Figure out how to use LOAD_CONTEXT_REG on GFX6-GFX7. */
device->has_load_ctx_reg_pkt = device->rad_info.chip_class >= GFX9 ||
- (device->rad_info.chip_class >= VI &&
+ (device->rad_info.chip_class >= GFX8 &&
device->rad_info.me_fw_feature >= 41);
@@ -770,5 +770,5 @@ void radv_GetPhysicalDeviceFeatures(
.shaderImageGatherExtended = true,
.shaderStorageImageExtendedFormats = true,
- .shaderStorageImageMultisample = pdevice->rad_info.chip_class >= VI,
+ .shaderStorageImageMultisample = pdevice->rad_info.chip_class >= GFX8,
.shaderUniformBufferArrayDynamicIndexing = true,
.shaderSampledImageArrayDynamicIndexing = true,
@@ -823,5 +823,5 @@ void radv_GetPhysicalDeviceFeatures2(
VkPhysicalDevice16BitStorageFeatures *features =
(VkPhysicalDevice16BitStorageFeatures*)ext;
- bool enabled = pdevice->rad_info.chip_class >= VI;
+ bool enabled = pdevice->rad_info.chip_class >= GFX8;
features->storageBuffer16BitAccess = enabled;
features->uniformAndStorageBuffer16BitAccess = enabled;
@@ -885,5 +885,5 @@ void radv_GetPhysicalDeviceFeatures2(
VkPhysicalDeviceScalarBlockLayoutFeaturesEXT *features =
(VkPhysicalDeviceScalarBlockLayoutFeaturesEXT *)ext;
- features->scalarBlockLayout = pdevice->rad_info.chip_class >= CIK;
+ features->scalarBlockLayout = pdevice->rad_info.chip_class >= GFX7;
break;
}
@@ -917,5 +917,5 @@ void radv_GetPhysicalDeviceFeatures2(
VkPhysicalDevice8BitStorageFeaturesKHR *features =
(VkPhysicalDevice8BitStorageFeaturesKHR*)ext;
- bool enabled = pdevice->rad_info.chip_class >= VI;
+ bool enabled = pdevice->rad_info.chip_class >= GFX8;
features->storageBuffer8BitAccess = enabled;
features->uniformAndStorageBuffer8BitAccess = enabled;
@@ -926,5 +926,5 @@ void radv_GetPhysicalDeviceFeatures2(
VkPhysicalDeviceFloat16Int8FeaturesKHR *features =
(VkPhysicalDeviceFloat16Int8FeaturesKHR*)ext;
- features->shaderFloat16 = pdevice->rad_info.chip_class >= VI && HAVE_LLVM >= 0x0800;
+ features->shaderFloat16 = pdevice->rad_info.chip_class >= GFX8 && HAVE_LLVM >= 0x0800;
features->shaderInt8 = true;
break;
@@ -1088,5 +1088,5 @@ void radv_GetPhysicalDeviceProperties(
.sampledImageDepthSampleCounts = sample_counts,
.sampledImageStencilSampleCounts = sample_counts,
- .storageImageSampleCounts = pdevice->rad_info.chip_class >= VI ? sample_counts : VK_SAMPLE_COUNT_1_BIT,
+ .storageImageSampleCounts = pdevice->rad_info.chip_class >= GFX8 ? sample_counts : VK_SAMPLE_COUNT_1_BIT,
.maxSampleMaskWords = 1,
.timestampComputeAndGraphics = true,
@@ -1177,5 +1177,5 @@ void radv_GetPhysicalDeviceProperties2(
VK_SUBGROUP_FEATURE_QUAD_BIT |
VK_SUBGROUP_FEATURE_VOTE_BIT;
- if (pdevice->rad_info.chip_class >= VI) {
+ if (pdevice->rad_info.chip_class >= GFX8) {
properties->supportedOperations |=
VK_SUBGROUP_FEATURE_ARITHMETIC_BIT |
@@ -1229,10 +1229,10 @@ void radv_GetPhysicalDeviceProperties2(
ac_get_num_physical_sgprs(pdevice->rad_info.chip_class);
properties->minSgprAllocation =
- pdevice->rad_info.chip_class >= VI ? 16 : 8;
+ pdevice->rad_info.chip_class >= GFX8 ? 16 : 8;
properties->maxSgprAllocation =
pdevice->rad_info.family == CHIP_TONGA ||
pdevice->rad_info.family == CHIP_ICELAND ? 96 : 104;
properties->sgprAllocationGranularity =
- pdevice->rad_info.chip_class >= VI ? 16 : 8;
+ pdevice->rad_info.chip_class >= GFX8 ? 16 : 8;
/* VGPR. */
@@ -1869,5 +1869,5 @@ VkResult radv_CreateDevice(
device->dispatch_initiator = S_00B800_COMPUTE_SHADER_EN(1);
- if (device->physical_device->rad_info.chip_class >= CIK) {
+ if (device->physical_device->rad_info.chip_class >= GFX7) {
/* If the KMD allows it (there is a KMD hw register for it),
* allow launching waves out-of-order.
@@ -1881,5 +1881,5 @@ VkResult radv_CreateDevice(
device->physical_device->rad_info.family == CHIP_HAWAII ? 4096 : 8192;
device->has_distributed_tess =
- device->physical_device->rad_info.chip_class >= VI &&
+ device->physical_device->rad_info.chip_class >= GFX8 &&
device->physical_device->rad_info.max_se >= 2;
@@ -1924,5 +1924,5 @@ VkResult radv_CreateDevice(
}
- if (device->physical_device->rad_info.chip_class >= CIK)
+ if (device->physical_device->rad_info.chip_class >= GFX7)
cik_create_gfx_config(device);
@@ -2224,5 +2224,5 @@ static unsigned
radv_get_hs_offchip_param(struct radv_device *device, uint32_t *max_offchip_buffers_p)
{
- bool double_offchip_buffers = device->physical_device->rad_info.chip_class >= CIK &&
+ bool double_offchip_buffers = device->physical_device->rad_info.chip_class >= GFX7 &&
device->physical_device->rad_info.family != CHIP_CARRIZO &&
device->physical_device->rad_info.family != CHIP_STONEY;
@@ -2235,5 +2235,5 @@ radv_get_hs_offchip_param(struct radv_device *device, uint32_t *max_offchip_buff
* Per RadeonSI:
* This must be one less than the maximum number due to a hw limitation.
- * Various hardware bugs in SI, CIK, and GFX9 need this.
+ * Various hardware bugs need thGFX7
*
* Per AMDVLK:
@@ -2245,6 +2245,6 @@ radv_get_hs_offchip_param(struct radv_device *device, uint32_t *max_offchip_buff
*/
if (device->physical_device->rad_info.family == CHIP_VEGA10 ||
- device->physical_device->rad_info.chip_class == CIK ||
- device->physical_device->rad_info.chip_class == SI)
+ device->physical_device->rad_info.chip_class == GFX7 ||
+ device->physical_device->rad_info.chip_class == GFX6)
--max_offchip_buffers_per_se;
@@ -2264,9 +2264,9 @@ radv_get_hs_offchip_param(struct radv_device *device, uint32_t *max_offchip_buff
switch (device->physical_device->rad_info.chip_class) {
- case SI:
+ case GFX6:
max_offchip_buffers = MIN2(max_offchip_buffers, 126);
break;
- case CIK:
- case VI:
+ case GFX7:
+ case GFX8:
case GFX9:
default:
@@ -2276,6 +2276,6 @@ radv_get_hs_offchip_param(struct radv_device *device, uint32_t *max_offchip_buff
*max_offchip_buffers_p = max_offchip_buffers;
- if (device->physical_device->rad_info.chip_class >= CIK) {
- if (device->physical_device->rad_info.chip_class >= VI)
+ if (device->physical_device->rad_info.chip_class >= GFX7) {
+ if (device->physical_device->rad_info.chip_class >= GFX8)
--max_offchip_buffers;
hs_offchip_param =
@@ -2305,5 +2305,5 @@ radv_emit_gs_ring_sizes(struct radv_queue *queue, struct radeon_cmdbuf *cs,
radv_cs_add_buffer(queue->device->ws, cs, gsvs_ring_bo);
- if (queue->device->physical_device->rad_info.chip_class >= CIK) {
+ if (queue->device->physical_device->rad_info.chip_class >= GFX7) {
radeon_set_uconfig_reg_seq(cs, R_030900_VGT_ESGS_RING_SIZE, 2);
radeon_emit(cs, esgs_ring_size >> 8);
@@ -2330,5 +2330,5 @@ radv_emit_tess_factor_ring(struct radv_queue *queue, struct radeon_cmdbuf *cs,
radv_cs_add_buffer(queue->device->ws, cs, tess_rings_bo);
- if (queue->device->physical_device->rad_info.chip_class >= CIK) {
+ if (queue->device->physical_device->rad_info.chip_class >= GFX7) {
radeon_set_uconfig_reg(cs, R_030938_VGT_TF_RING_SIZE,
S_030938_SIZE(tf_ring_size / 4));
@@ -2650,5 +2650,5 @@ radv_get_preamble_cs(struct radv_queue *queue,
NULL, 0,
queue->queue_family_index == RING_COMPUTE &&
- queue->device->physical_device->rad_info.chip_class >= CIK,
+ queue->device->physical_device->rad_info.chip_class >= GFX7,
(queue->queue_family_index == RADV_QUEUE_COMPUTE ? RADV_CMD_FLAG_CS_PARTIAL_FLUSH : (RADV_CMD_FLAG_CS_PARTIAL_FLUSH | RADV_CMD_FLAG_PS_PARTIAL_FLUSH)) |
RADV_CMD_FLAG_INV_ICACHE |
@@ -2662,5 +2662,5 @@ radv_get_preamble_cs(struct radv_queue *queue,
NULL, 0,
queue->queue_family_index == RING_COMPUTE &&
- queue->device->physical_device->rad_info.chip_class >= CIK,
+ queue->device->physical_device->rad_info.chip_class >= GFX7,
RADV_CMD_FLAG_INV_ICACHE |
RADV_CMD_FLAG_INV_SMEM_L1 |
@@ -4275,5 +4275,5 @@ radv_initialise_color_surface(struct radv_device *device,
if (radv_image_has_fmask(iview->image)) {
- if (device->physical_device->rad_info.chip_class >= CIK)
+ if (device->physical_device->rad_info.chip_class >= GFX7)
cb->cb_color_pitch |= S_028C64_FMASK_TILE_MAX(iview->image->fmask.pitch_in_pixels / 8 - 1);
cb->cb_color_attrib |= S_028C74_FMASK_TILE_MODE_INDEX(iview->image->fmask.tile_mode_index);
@@ -4281,5 +4281,5 @@ radv_initialise_color_surface(struct radv_device *device,
} else {
/* This must be set for fast clear to work without FMASK. */
- if (device->physical_device->rad_info.chip_class >= CIK)
+ if (device->physical_device->rad_info.chip_class >= GFX7)
cb->cb_color_pitch |= S_028C64_FMASK_TILE_MAX(pitch_tile_max);
cb->cb_color_attrib |= S_028C74_FMASK_TILE_MODE_INDEX(tile_mode_index);
@@ -4361,5 +4361,5 @@ radv_initialise_color_surface(struct radv_device *device,
if (radv_image_has_fmask(iview->image)) {
cb->cb_color_info |= S_028C70_COMPRESSION(1);
- if (device->physical_device->rad_info.chip_class == SI) {
+ if (device->physical_device->rad_info.chip_class == GFX6) {
unsigned fmask_bankh = util_logbase2(iview->image->fmask.bank_height);
cb->cb_color_attrib |= S_028C74_FMASK_BANK_HEIGHT(fmask_bankh);
@@ -4378,5 +4378,5 @@ radv_initialise_color_surface(struct radv_device *device,
/* This must be set for fast clear to work without FMASK. */
if (!radv_image_has_fmask(iview->image) &&
- device->physical_device->rad_info.chip_class == SI) {
+ device->physical_device->rad_info.chip_class == GFX6) {
unsigned bankh = util_logbase2(surf->u.legacy.bankh);
cb->cb_color_attrib |= S_028C74_FMASK_BANK_HEIGHT(bankh);
@@ -4549,5 +4549,5 @@ radv_initialise_ds_surface(struct radv_device *device,
ds->db_z_info |= S_028040_NUM_SAMPLES(util_logbase2(iview->image->info.samples));
- if (device->physical_device->rad_info.chip_class >= CIK) {
+ if (device->physical_device->rad_info.chip_class >= GFX7) {
struct radeon_info *info = &device->physical_device->rad_info;
unsigned tiling_index = surf->u.legacy.tiling_index[level];
@@ -4808,5 +4808,5 @@ radv_init_sampler(struct radv_device *device,
uint32_t max_aniso = radv_get_max_anisotropy(device, pCreateInfo);
uint32_t max_aniso_ratio = radv_tex_aniso_filter(max_aniso);
- bool is_vi = (device->physical_device->rad_info.chip_class >= VI);
+ bool is_vi = (device->physical_device->rad_info.chip_class >= GFX8);
unsigned filter_mode = V_008F30_SQ_IMG_FILTER_MODE_BLEND;
@@ -4836,5 +4836,5 @@ radv_init_sampler(struct radv_device *device,
S_008F38_MIP_FILTER(radv_tex_mipfilter(pCreateInfo->mipmapMode)) |
S_008F38_MIP_POINT_PRECLAMP(0) |
- S_008F38_DISABLE_LSB_CEIL(device->physical_device->rad_info.chip_class <= VI) |
+ S_008F38_DISABLE_LSB_CEIL(device->physical_device->rad_info.chip_class <= GFX8) |
S_008F38_FILTER_PREC_FIX(1) |
S_008F38_ANISO_OVERRIDE(is_vi));
diff --git a/src/amd/vulkan/radv_extensions.py b/src/amd/vulkan/radv_extensions.py
index 576a21f4ca5..0b5af56a435 100644
--- a/src/amd/vulkan/radv_extensions.py
+++ b/src/amd/vulkan/radv_extensions.py
@@ -97,5 +97,5 @@ EXTENSIONS = [
Extension('VK_KHR_multiview', 1, True),
Extension('VK_KHR_display', 23, 'VK_USE_PLATFORM_DISPLAY_KHR'),
- Extension('VK_KHR_8bit_storage', 1, 'device->rad_info.chip_class >= VI'),
+ Extension('VK_KHR_8bit_storage', 1, 'device->rad_info.chip_class >= GFX8'),
Extension('VK_EXT_direct_mode_display', 1, 'VK_USE_PLATFORM_DISPLAY_KHR'),
Extension('VK_EXT_acquire_xlib_display', 1, 'VK_USE_PLATFORM_XLIB_XRANDR_EXT'),
@@ -120,6 +120,6 @@ EXTENSIONS = [
Extension('VK_EXT_pci_bus_info', 2, True),
Extension('VK_EXT_pipeline_creation_feedback', 1, True),
- Extension('VK_EXT_sampler_filter_minmax', 1, 'device->rad_info.chip_class >= CIK'),
- Extension('VK_EXT_scalar_block_layout', 1, 'device->rad_info.chip_class >= CIK'),
+ Extension('VK_EXT_sampler_filter_minmax', 1, 'device->rad_info.chip_class >= GFX7'),
+ Extension('VK_EXT_scalar_block_layout', 1, 'device->rad_info.chip_class >= GFX7'),
Extension('VK_EXT_shader_viewport_index_layer', 1, True),
Extension('VK_EXT_shader_stencil_export', 1, True),
@@ -129,6 +129,6 @@ EXTENSIONS = [
Extension('VK_AMD_draw_indirect_count', 1, True),
Extension('VK_AMD_gcn_shader', 1, True),
- Extension('VK_AMD_gpu_shader_half_float', 1, 'device->rad_info.chip_class >= VI && HAVE_LLVM >= 0x0800'),
- Extension('VK_AMD_gpu_shader_int16', 1, 'device->rad_info.chip_class >= VI'),
+ Extension('VK_AMD_gpu_shader_half_float', 1, 'device->rad_info.chip_class >= GFX8 && HAVE_LLVM >= 0x0800'),
+ Extension('VK_AMD_gpu_shader_int16', 1, 'device->rad_info.chip_class >= GFX8'),
Extension('VK_AMD_rasterization_order', 1, 'device->has_out_of_order_rast'),
Extension('VK_AMD_shader_core_properties', 1, True),
@@ -137,5 +137,5 @@ EXTENSIONS = [
Extension('VK_GOOGLE_decorate_string', 1, True),
Extension('VK_GOOGLE_hlsl_functionality1', 1, True),
- Extension('VK_NV_compute_shader_derivatives', 1, 'device->rad_info.chip_class >= VI'),
+ Extension('VK_NV_compute_shader_derivatives', 1, 'device->rad_info.chip_class >= GFX8'),
]
diff --git a/src/amd/vulkan/radv_formats.c b/src/amd/vulkan/radv_formats.c
index 9883002fa42..d7b560082f6 100644
--- a/src/amd/vulkan/radv_formats.c
+++ b/src/amd/vulkan/radv_formats.c
@@ -762,5 +762,5 @@ radv_physical_device_get_format_properties(struct radv_physical_device *physical
case VK_FORMAT_A2R10G10B10_SINT_PACK32:
case VK_FORMAT_A2B10G10R10_SINT_PACK32:
- if (physical_device->rad_info.chip_class <= VI &&
+ if (physical_device->rad_info.chip_class <= GFX8 &&
physical_device->rad_info.family != CHIP_STONEY) {
buffer &= ~(VK_FORMAT_FEATURE_UNIFORM_TEXEL_BUFFER_BIT |
diff --git a/src/amd/vulkan/radv_image.c b/src/amd/vulkan/radv_image.c
index 3ffb4e95749..161997ae196 100644
--- a/src/amd/vulkan/radv_image.c
+++ b/src/amd/vulkan/radv_image.c
@@ -48,5 +48,5 @@ radv_choose_tiling(struct radv_device *device,
if (!vk_format_is_compressed(pCreateInfo->format) &&
!vk_format_is_depth_or_stencil(pCreateInfo->format)
- && device->physical_device->rad_info.chip_class <= VI) {
+ && device->physical_device->rad_info.chip_class <= GFX8) {
/* this causes hangs in some VK CTS tests on GFX9. */
/* Textures with a very small height are recommended to be linear. */
@@ -70,5 +70,5 @@ radv_use_tc_compat_htile_for_image(struct radv_device *device,
{
/* TC-compat HTILE is only available for GFX8+. */
- if (device->physical_device->rad_info.chip_class < VI)
+ if (device->physical_device->rad_info.chip_class < GFX8)
return false;
@@ -131,5 +131,5 @@ radv_use_dcc_for_image(struct radv_device *device,
/* DCC (Delta Color Compression) is only available for GFX8+. */
- if (device->physical_device->rad_info.chip_class < VI)
+ if (device->physical_device->rad_info.chip_class < GFX8)
return false;
@@ -329,5 +329,5 @@ radv_make_buffer_descriptor(struct radv_device *device,
S_008F04_STRIDE(stride);
- if (device->physical_device->rad_info.chip_class != VI && stride) {
+ if (device->physical_device->rad_info.chip_class != GFX8 && stride) {
range /= stride;
}
@@ -371,10 +371,10 @@ si_set_mutable_tex_desc_fields(struct radv_device *device,
state[1] |= S_008F14_BASE_ADDRESS_HI(va >> 40);
- if (chip_class >= VI) {
+ if (chip_class >= GFX8) {
state[6] &= C_008F28_COMPRESSION_EN;
state[7] = 0;
if (!is_storage_image && radv_dcc_enabled(image, first_level)) {
meta_va = gpu_address + image->dcc_offset;
- if (chip_class <= VI)
+ if (chip_class <= GFX8)
meta_va += base_level_info->dcc_offset;
} else if (!is_storage_image &&
@@ -418,5 +418,5 @@ si_set_mutable_tex_desc_fields(struct radv_device *device,
}
} else {
- /* SI-CI-VI */
+ /* GFX6-GFX8 */
unsigned pitch = base_level_info->nblk_x * block_width;
unsigned index = si_tile_mode_index(plane, base_level, is_stencil);
@@ -597,5 +597,5 @@ si_make_texture_descriptor(struct radv_device *device,
* bits in the first dword of sampler state.
*/
- if (device->physical_device->rad_info.chip_class <= CIK && image->info.samples <= 1) {
+ if (device->physical_device->rad_info.chip_class <= GFX7 && image->info.samples <= 1) {
if (first_level == last_level)
state[7] = C_008F30_MAX_ANISO_RATIO;
@@ -726,5 +726,5 @@ radv_query_opaque_metadata(struct radv_device *device,
/* Dwords [10:..] contain the mipmap level offsets. */
- if (device->physical_device->rad_info.chip_class <= VI) {
+ if (device->physical_device->rad_info.chip_class <= GFX8) {
for (i = 0; i <= image->info.levels - 1; i++)
md->metadata[10+i] = image->planes[0].surface.u.legacy.level[i].offset >> 8;
diff --git a/src/amd/vulkan/radv_nir_to_llvm.c b/src/amd/vulkan/radv_nir_to_llvm.c
index e8be058d3f7..341f6388f32 100644
--- a/src/amd/vulkan/radv_nir_to_llvm.c
+++ b/src/amd/vulkan/radv_nir_to_llvm.c
@@ -263,5 +263,5 @@ get_tcs_num_patches(struct radv_shader_context *ctx)
* Test: dEQP-VK.tessellation.shader_input_output.barrier
*/
- if (ctx->options->chip_class >= CIK && ctx->options->family != CHIP_STONEY)
+ if (ctx->options->chip_class >= GFX7 && ctx->options->family != CHIP_STONEY)
hardware_lds_size = 65536;
@@ -274,6 +274,6 @@ get_tcs_num_patches(struct radv_shader_context *ctx)
num_patches = MIN2(num_patches, 40);
- /* SI bug workaround - limit LS-HS threadgroups to only one wave. */
- if (ctx->options->chip_class == SI) {
+ /* GFX6 bug workaround - limit LS-HS threadgroups to only one wave. */
+ if (ctx->options->chip_class == GFX6) {
unsigned one_wave = 64 / MAX2(num_tcs_input_cp, num_tcs_output_cp);
num_patches = MIN2(num_patches, one_wave);
@@ -3277,5 +3277,5 @@ write_tess_factors(struct radv_shader_context *ctx)
unsigned tf_offset = 0;
- if (ctx->options->chip_class <= VI) {
+ if (ctx->options->chip_class <= GFX8) {
ac_nir_build_if(&inner_if_ctx, ctx,
LLVMBuildICmp(ctx->ac.builder, LLVMIntEQ,
@@ -3519,5 +3519,5 @@ static void
ac_setup_rings(struct radv_shader_context *ctx)
{
- if (ctx->options->chip_class <= VI &&
+ if (ctx->options->chip_class <= GFX8 &&
(ctx->stage == MESA_SHADER_GEOMETRY ||
ctx->options->key.vs.as_es || ctx->options->key.tes.as_es)) {
@@ -3569,5 +3569,5 @@ ac_setup_rings(struct radv_shader_context *ctx)
stride = 4 * num_components * ctx->gs_max_out_vertices;
- /* Limit on the stride field for <= CIK. */
+ /* Limit on the stride field for <= GFX7. */
assert(stride < (1 << 14));
@@ -3617,5 +3617,5 @@ radv_nir_get_max_workgroup_size(enum chip_class chip_class,
switch (nir->info.stage) {
case MESA_SHADER_TESS_CTRL:
- return chip_class >= CIK ? 128 : 64;
+ return chip_class >= GFX7 ? 128 : 64;
case MESA_SHADER_GEOMETRY:
return chip_class >= GFX9 ? 128 : 64;
@@ -3962,5 +3962,5 @@ static void ac_compile_llvm_module(struct ac_llvm_compiler *ac_llvm,
* - Some opcodes don't support denormals, such as v_mad_f32. We would
* have to stop using those.
- * - SI & CI would be very slow.
+ * - GFX6 & GFX7 would be very slow.
*/
config->float_mode |= V_00B028_FP_64_DENORMS;
diff --git a/src/amd/vulkan/radv_pipeline.c b/src/amd/vulkan/radv_pipeline.c
index f25a5f55bf5..c89a6f139ba 100644
--- a/src/amd/vulkan/radv_pipeline.c
+++ b/src/amd/vulkan/radv_pipeline.c
@@ -1559,9 +1559,9 @@ calculate_gs_ring_sizes(struct radv_pipeline *pipeline, const struct radv_gs_sta
unsigned wave_size = 64;
unsigned max_gs_waves = 32 * num_se; /* max 32 per SE on GCN */
- /* On SI-CI, the value comes from VGT_GS_VERTEX_REUSE = 16.
- * On VI+, the value comes from VGT_VERTEX_REUSE_BLOCK_CNTL = 30 (+2).
+ /* On GFX6-GFX7, the value comes from VGT_GS_VERTEX_REUSE = 16.
+ * On GFX8+, the value comes from VGT_VERTEX_REUSE_BLOCK_CNTL = 30 (+2).
*/
unsigned gs_vertex_reuse =
- (device->physical_device->rad_info.chip_class >= VI ? 32 : 16) * num_se;
+ (device->physical_device->rad_info.chip_class >= GFX8 ? 32 : 16) * num_se;
unsigned alignment = 256 * num_se;
/* The maximum size is 63.999 MB per SE. */
@@ -1582,5 +1582,5 @@ calculate_gs_ring_sizes(struct radv_pipeline *pipeline, const struct radv_gs_sta
gsvs_ring_size = align(gsvs_ring_size, alignment);
- if (pipeline->device->physical_device->rad_info.chip_class <= VI)
+ if (pipeline->device->physical_device->rad_info.chip_class <= GFX8)
pipeline->graphics.esgs_ring_size = CLAMP(esgs_ring_size, min_esgs_ring_size, max_size);
@@ -1644,5 +1644,5 @@ calculate_tess_state(struct radv_pipeline *pipeline,
lds_size = pipeline->shaders[MESA_SHADER_TESS_CTRL]->info.tcs.lds_size;
- if (pipeline->device->physical_device->rad_info.chip_class >= CIK) {
+ if (pipeline->device->physical_device->rad_info.chip_class >= GFX7) {
assert(lds_size <= 65536);
lds_size = align(lds_size, 512) / 512;
@@ -1905,5 +1905,5 @@ radv_generate_graphics_pipeline_key(struct radv_pipeline *pipeline,
key.vertex_attribute_strides[location] = radv_get_attrib_stride(input_state, desc->binding);
- if (pipeline->device->physical_device->rad_info.chip_class <= VI &&
+ if (pipeline->device->physical_device->rad_info.chip_class <= GFX8 &&
pipeline->device->physical_device->rad_info.family != CHIP_STONEY) {
VkFormat format = input_state->pVertexAttributeDescriptions[i].format;
@@ -1963,5 +1963,5 @@ radv_generate_graphics_pipeline_key(struct radv_pipeline *pipeline,
key.col_format = blend->spi_shader_col_format;
- if (pipeline->device->physical_device->rad_info.chip_class < VI)
+ if (pipeline->device->physical_device->rad_info.chip_class < GFX8)
radv_pipeline_compute_get_int_clamp(pCreateInfo, &key.is_int8, &key.is_int10);
@@ -2919,5 +2919,5 @@ radv_pipeline_generate_multisample_state(struct radeon_cmdbuf *ctx_cs,
* currently always TRUE because the driver doesn't support 16 samples.
*/
- bool exclusion = pipeline->device->physical_device->rad_info.chip_class >= CIK;
+ bool exclusion = pipeline->device->physical_device->rad_info.chip_class >= GFX7;
radeon_set_context_reg(ctx_cs, R_02882C_PA_SU_PRIM_FILTER_CNTL,
S_02882C_XMAX_RIGHT_EXCLUSION(exclusion) |
@@ -3004,5 +3004,5 @@ radv_pipeline_generate_hw_vs(struct radeon_cmdbuf *ctx_cs,
clip_dist_mask);
- if (pipeline->device->physical_device->rad_info.chip_class <= VI)
+ if (pipeline->device->physical_device->rad_info.chip_class <= GFX8)
radeon_set_context_reg(ctx_cs, R_028AB4_VGT_REUSE_OFF,
outinfo->writes_viewport_index);
@@ -3037,5 +3037,5 @@ radv_pipeline_generate_hw_ls(struct radeon_cmdbuf *cs,
rsrc2 |= S_00B52C_LDS_SIZE(tess->lds_size);
- if (pipeline->device->physical_device->rad_info.chip_class == CIK &&
+ if (pipeline->device->physical_device->rad_info.chip_class == GFX7 &&
pipeline->device->physical_device->rad_info.family != CHIP_HAWAII)
radeon_set_sh_reg(cs, R_00B52C_SPI_SHADER_PGM_RSRC2_LS, rsrc2);
@@ -3119,5 +3119,5 @@ radv_pipeline_generate_tess_shaders(struct radeon_cmdbuf *ctx_cs,
tess->tf_param);
- if (pipeline->device->physical_device->rad_info.chip_class >= CIK)
+ if (pipeline->device->physical_device->rad_info.chip_class >= GFX7)
radeon_set_context_reg_idx(ctx_cs, R_028B58_VGT_LS_HS_CONFIG, 2,
tess->ls_hs_config);
@@ -3491,5 +3491,5 @@ radv_pipeline_generate_pm4(struct radv_pipeline *pipeline,
radeon_set_context_reg(ctx_cs, R_028B54_VGT_SHADER_STAGES_EN, radv_compute_vgt_shader_stages_en(pipeline));
- if (pipeline->device->physical_device->rad_info.chip_class >= CIK) {
+ if (pipeline->device->physical_device->rad_info.chip_class >= GFX7) {
radeon_set_uconfig_reg_idx(cs, R_030908_VGT_PRIMITIVE_TYPE, 1, prim);
} else {
@@ -3523,10 +3523,10 @@ radv_compute_ia_multi_vgt_param_helpers(struct radv_pipeline *pipeline,
/* GS requirement. */
ia_multi_vgt_param.partial_es_wave = false;
- if (radv_pipeline_has_gs(pipeline) && device->physical_device->rad_info.chip_class <= VI)
+ if (radv_pipeline_has_gs(pipeline) && device->physical_device->rad_info.chip_class <= GFX8)
if (SI_GS_PER_ES / ia_multi_vgt_param.primgroup_size >= pipeline->device->gs_table_depth - 3)
ia_multi_vgt_param.partial_es_wave = true;
ia_multi_vgt_param.wd_switch_on_eop = false;
- if (device->physical_device->rad_info.chip_class >= CIK) {
+ if (device->physical_device->rad_info.chip_class >= GFX7) {
/* WD_SWITCH_ON_EOP has no effect on GPUs with less than
* 4 shader engines. Set 1 to pass the assertion below.
@@ -3568,5 +3568,5 @@ radv_compute_ia_multi_vgt_param_helpers(struct radv_pipeline *pipeline,
if (device->has_distributed_tess) {
if (radv_pipeline_has_gs(pipeline)) {
- if (device->physical_device->rad_info.chip_class <= VI)
+ if (device->physical_device->rad_info.chip_class <= GFX8)
ia_multi_vgt_param.partial_es_wave = true;
} else {
@@ -3610,5 +3610,5 @@ radv_compute_ia_multi_vgt_param_helpers(struct radv_pipeline *pipeline,
S_028AA8_PRIMGROUP_SIZE(ia_multi_vgt_param.primgroup_size - 1) |
/* The following field was moved to VGT_SHADER_STAGES_EN in GFX9. */
- S_028AA8_MAX_PRIMGRP_IN_WAVE(device->physical_device->rad_info.chip_class == VI ? 2 : 0) |
+ S_028AA8_MAX_PRIMGRP_IN_WAVE(device->physical_device->rad_info.chip_class == GFX8 ? 2 : 0) |
S_030960_EN_INST_OPT_BASIC(device->physical_device->rad_info.chip_class >= GFX9) |
S_030960_EN_INST_OPT_ADV(device->physical_device->rad_info.chip_class >= GFX9);
@@ -3886,5 +3886,5 @@ radv_compute_generate_pm4(struct radv_pipeline *pipeline)
S_00B854_SIMD_DEST_CNTL(waves_per_threadgroup % 4 == 0);
- if (device->physical_device->rad_info.chip_class >= CIK) {
+ if (device->physical_device->rad_info.chip_class >= GFX7) {
unsigned num_cu_per_se =
device->physical_device->rad_info.num_good_compute_units /
diff --git a/src/amd/vulkan/radv_private.h b/src/amd/vulkan/radv_private.h
index aa25e8f9805..a88c0f31ad3 100644
--- a/src/amd/vulkan/radv_private.h
+++ b/src/amd/vulkan/radv_private.h
@@ -704,5 +704,5 @@ struct radv_device {
float sample_locations_16x[16][2];
- /* CIK and later */
+ /* GFX7 and later */
uint32_t gfx_init_size_dw;
struct radeon_winsys_bo *gfx_init;
diff --git a/src/amd/vulkan/radv_shader.c b/src/amd/vulkan/radv_shader.c
index 17d6c5bc33a..dfa50155c06 100644
--- a/src/amd/vulkan/radv_shader.c
+++ b/src/amd/vulkan/radv_shader.c
@@ -774,5 +774,5 @@ generate_shader_stats(struct radv_device *device,
{
enum chip_class chip_class = device->physical_device->rad_info.chip_class;
- unsigned lds_increment = chip_class >= CIK ? 512 : 256;
+ unsigned lds_increment = chip_class >= GFX7 ? 512 : 256;
struct ac_shader_config *conf;
unsigned max_simd_waves;
@@ -876,5 +876,5 @@ radv_GetShaderInfoAMD(VkDevice _device,
*pInfoSize = sizeof(VkShaderStatisticsInfoAMD);
} else {
- unsigned lds_multiplier = device->physical_device->rad_info.chip_class >= CIK ? 512 : 256;
+ unsigned lds_multiplier = device->physical_device->rad_info.chip_class >= GFX7 ? 512 : 256;
struct ac_shader_config *conf = &variant->config;
diff --git a/src/amd/vulkan/si_cmd_buffer.c b/src/amd/vulkan/si_cmd_buffer.c
index e73c13762e5..0f4bdadc3d2 100644
--- a/src/amd/vulkan/si_cmd_buffer.c
+++ b/src/amd/vulkan/si_cmd_buffer.c
@@ -26,5 +26,5 @@
*/
-/* command buffer handling for SI */
+/* command buffer handling for AMD GCN */
#include "radv_private.h"
@@ -52,6 +52,6 @@ si_write_harvested_raster_configs(struct radv_physical_device *physical_device,
for (se = 0; se < num_se; se++) {
- /* GRBM_GFX_INDEX has a different offset on SI and CI+ */
- if (physical_device->rad_info.chip_class < CIK)
+ /* GRBM_GFX_INDEX has a different offset on GFX6 and GFX7+ */
+ if (physical_device->rad_info.chip_class < GFX7)
radeon_set_config_reg(cs, R_00802C_GRBM_GFX_INDEX,
S_00802C_SE_INDEX(se) |
@@ -65,6 +65,6 @@ si_write_harvested_raster_configs(struct radv_physical_device *physical_device,
}
- /* GRBM_GFX_INDEX has a different offset on SI and CI+ */
- if (physical_device->rad_info.chip_class < CIK)
+ /* GRBM_GFX_INDEX has a different offset on GFX6 and GFX7+ */
+ if (physical_device->rad_info.chip_class < GFX7)
radeon_set_config_reg(cs, R_00802C_GRBM_GFX_INDEX,
S_00802C_SE_BROADCAST_WRITES(1) |
@@ -76,5 +76,5 @@ si_write_harvested_raster_configs(struct radv_physical_device *physical_device,
S_030800_INSTANCE_BROADCAST_WRITES(1));
- if (physical_device->rad_info.chip_class >= CIK)
+ if (physical_device->rad_info.chip_class >= GFX7)
radeon_set_context_reg(cs, R_028354_PA_SC_RASTER_CONFIG_1, raster_config_1);
}
@@ -94,5 +94,5 @@ si_emit_compute(struct radv_physical_device *physical_device,
radeon_emit(cs, S_00B85C_SH0_CU_EN(0xffff) | S_00B85C_SH1_CU_EN(0xffff));
- if (physical_device->rad_info.chip_class >= CIK) {
+ if (physical_device->rad_info.chip_class >= GFX7) {
/* Also set R_00B858_COMPUTE_STATIC_THREAD_MGMT_SE2 / SE3 */
radeon_set_sh_reg_seq(cs,
@@ -109,5 +109,5 @@ si_emit_compute(struct radv_physical_device *physical_device,
* which is now 0x22f.
*/
- if (physical_device->rad_info.chip_class <= SI) {
+ if (physical_device->rad_info.chip_class <= GFX6) {
/* XXX: This should be:
* (number of compute units) * 4 * (waves per simd) - 1 */
@@ -143,5 +143,5 @@ si_set_raster_config(struct radv_physical_device *physical_device,
radeon_set_context_reg(cs, R_028350_PA_SC_RASTER_CONFIG,
raster_config);
- if (physical_device->rad_info.chip_class >= CIK)
+ if (physical_device->rad_info.chip_class >= GFX7)
radeon_set_context_reg(cs, R_028354_PA_SC_RASTER_CONFIG_1,
raster_config_1);
@@ -159,7 +159,7 @@ si_emit_graphics(struct radv_physical_device *physical_device,
int i;
- /* Only SI can disable CLEAR_STATE for now. */
+ /* Only GFX6 can disable CLEAR_STATE for now. */
assert(physical_device->has_clear_state ||
- physical_device->rad_info.chip_class == SI);
+ physical_device->rad_info.chip_class == GFX6);
radeon_emit(cs, PKT3(PKT3_CONTEXT_CONTROL, 1, 0));
@@ -172,5 +172,5 @@ si_emit_graphics(struct radv_physical_device *physical_device,
}
- if (physical_device->rad_info.chip_class <= VI)
+ if (physical_device->rad_info.chip_class <= GFX8)
si_set_raster_config(physical_device, cs);
@@ -180,5 +180,5 @@ si_emit_graphics(struct radv_physical_device *physical_device,
/* FIXME calculate these values somehow ??? */
- if (physical_device->rad_info.chip_class <= VI) {
+ if (physical_device->rad_info.chip_class <= GFX8) {
radeon_set_context_reg(cs, R_028A54_VGT_GS_PER_ES, SI_GS_PER_ES);
radeon_set_context_reg(cs, R_028A58_VGT_ES_PER_GS, 0x40);
@@ -194,5 +194,5 @@ si_emit_graphics(struct radv_physical_device *physical_device,
if (!physical_device->has_clear_state)
radeon_set_context_reg(cs, R_028AB8_VGT_VTX_CNT_EN, 0x0);
- if (physical_device->rad_info.chip_class < CIK)
+ if (physical_device->rad_info.chip_class < GFX7)
radeon_set_config_reg(cs, R_008A14_PA_CL_ENHANCE, S_008A14_NUM_CLIP_SEQ(3) |
S_008A14_CLIP_VTX_REORDER_ENA(1));
@@ -207,5 +207,5 @@ si_emit_graphics(struct radv_physical_device *physical_device,
* I don't know why. Deduced by trial and error.
*/
- if (physical_device->rad_info.chip_class <= CIK) {
+ if (physical_device->rad_info.chip_class <= GFX7) {
radeon_set_context_reg(cs, R_028B28_VGT_STRMOUT_DRAW_OPAQUE_OFFSET, 0);
radeon_set_context_reg(cs, R_028204_PA_SC_WINDOW_SCISSOR_TL,
@@ -230,5 +230,5 @@ si_emit_graphics(struct radv_physical_device *physical_device,
radeon_set_context_reg(cs, R_02820C_PA_SC_CLIPRECT_RULE, 0xFFFF);
radeon_set_context_reg(cs, R_028230_PA_SC_EDGERULE, 0xAAAAAAAA);
- /* PA_SU_HARDWARE_SCREEN_OFFSET must be 0 due to hw bug on SI */
+ /* PA_SU_HARDWARE_SCREEN_OFFSET must be 0 due to hw bug on GFX6 */
radeon_set_context_reg(cs, R_028234_PA_SU_HARDWARE_SCREEN_OFFSET, 0);
radeon_set_context_reg(cs, R_028820_PA_CL_NANINF_CNTL, 0);
@@ -257,5 +257,5 @@ si_emit_graphics(struct radv_physical_device *physical_device,
}
- if (physical_device->rad_info.chip_class >= CIK) {
+ if (physical_device->rad_info.chip_class >= GFX7) {
if (physical_device->rad_info.chip_class >= GFX9) {
radeon_set_sh_reg(cs, R_00B41C_SPI_SHADER_PGM_RSRC3_HS,
@@ -304,5 +304,5 @@ si_emit_graphics(struct radv_physical_device *physical_device,
}
- if (physical_device->rad_info.chip_class >= VI) {
+ if (physical_device->rad_info.chip_class >= GFX8) {
uint32_t vgt_tess_distribution;
@@ -587,5 +587,5 @@ si_get_ia_multi_vgt_param(struct radv_cmd_buffer *cmd_buffer,
partial_vs_wave = cmd_buffer->state.pipeline->graphics.ia_multi_vgt_param.partial_vs_wave;
- if (chip_class >= CIK) {
+ if (chip_class >= GFX7) {
wd_switch_on_eop = cmd_buffer->state.pipeline->graphics.ia_multi_vgt_param.wd_switch_on_eop;
@@ -602,17 +602,17 @@ si_get_ia_multi_vgt_param(struct radv_cmd_buffer *cmd_buffer,
* This is needed for good VS wave utilization.
*/
- if (chip_class <= VI &&
+ if (chip_class <= GFX8 &&
info->max_se == 4 &&
multi_instances_smaller_than_primgroup)
wd_switch_on_eop = true;
- /* Required on CIK and later. */
+ /* Required on GFX7 and later. */
if (info->max_se > 2 && !wd_switch_on_eop)
ia_switch_on_eoi = true;
- /* Required by Hawaii and, for some special cases, by VI. */
+ /* Required by Hawaii and, for some special cases, by GFX8. */
if (ia_switch_on_eoi &&
(family == CHIP_HAWAII ||
- (chip_class == VI &&
+ (chip_class == GFX8 &&
/* max primgroup in wave is always 2 - leave this for documentation */
(radv_pipeline_has_gs(cmd_buffer->state.pipeline) || max_primgroup_in_wave != 2))))
@@ -634,5 +634,5 @@ si_get_ia_multi_vgt_param(struct radv_cmd_buffer *cmd_buffer,
}
/* If SWITCH_ON_EOI is set, PARTIAL_ES_WAVE must be set too. */
- if (chip_class <= VI && ia_switch_on_eoi)
+ if (chip_class <= GFX8 && ia_switch_on_eoi)
partial_es_wave = true;
@@ -659,5 +659,5 @@ si_get_ia_multi_vgt_param(struct radv_cmd_buffer *cmd_buffer,
S_028AA8_PARTIAL_VS_WAVE_ON(partial_vs_wave) |
S_028AA8_PARTIAL_ES_WAVE_ON(partial_es_wave) |
- S_028AA8_WD_SWITCH_ON_EOP(chip_class >= CIK ? wd_switch_on_eop : 0);
+ S_028AA8_WD_SWITCH_ON_EOP(chip_class >= GFX7 ? wd_switch_on_eop : 0);
}
@@ -705,6 +705,6 @@ void si_cs_emit_write_event_eop(struct radeon_cmdbuf *cs,
radeon_emit(cs, 0); /* unused */
} else {
- if (chip_class == CIK ||
- chip_class == VI) {
+ if (chip_class == GFX7 ||
+ chip_class == GFX8) {
/* Two EOP events are required to make all engines go idle
* (and optional cache flushes executed) before the timestamp
@@ -789,5 +789,5 @@ si_cs_emit_cache_flush(struct radeon_cmdbuf *cs,
cp_coher_cntl |= S_0085F0_SH_KCACHE_ACTION_ENA(1);
- if (chip_class <= VI) {
+ if (chip_class <= GFX8) {
if (flush_bits & RADV_CMD_FLAG_FLUSH_AND_INV_CB) {
cp_coher_cntl |= S_0085F0_CB_ACTION_ENA(1) |
@@ -802,5 +802,5 @@ si_cs_emit_cache_flush(struct radeon_cmdbuf *cs,
/* Necessary for DCC */
- if (chip_class >= VI) {
+ if (chip_class >= GFX8) {
si_cs_emit_write_event_eop(cs,
chip_class,
@@ -912,10 +912,10 @@ si_cs_emit_cache_flush(struct radeon_cmdbuf *cs,
if ((flush_bits & RADV_CMD_FLAG_INV_GLOBAL_L2) ||
- (chip_class <= CIK && (flush_bits & RADV_CMD_FLAG_WRITEBACK_GLOBAL_L2))) {
+ (chip_class <= GFX7 && (flush_bits & RADV_CMD_FLAG_WRITEBACK_GLOBAL_L2))) {
si_emit_acquire_mem(cs, is_mec, chip_class >= GFX9,
cp_coher_cntl |
S_0085F0_TC_ACTION_ENA(1) |
S_0085F0_TCL1_ACTION_ENA(1) |
- S_0301F0_TC_WB_ACTION_ENA(chip_class >= VI));
+ S_0301F0_TC_WB_ACTION_ENA(chip_class >= GFX8));
cp_coher_cntl = 0;
} else {
@@ -1100,5 +1100,5 @@ static void si_emit_cp_dma(struct radv_cmd_buffer *cmd_buffer,
header |= S_411_SRC_SEL(V_411_SRC_ADDR_TC_L2);
- if (cmd_buffer->device->physical_device->rad_info.chip_class >= CIK) {
+ if (cmd_buffer->device->physical_device->rad_info.chip_class >= GFX7) {
radeon_emit(cs, PKT3(PKT3_DMA_DATA, 5, cmd_buffer->state.predicating));
radeon_emit(cs, header);
@@ -1282,5 +1282,5 @@ void si_cp_dma_clear_buffer(struct radv_cmd_buffer *cmd_buffer, uint64_t va,
void si_cp_dma_wait_for_idle(struct radv_cmd_buffer *cmd_buffer)
{
- if (cmd_buffer->device->physical_device->rad_info.chip_class < CIK)
+ if (cmd_buffer->device->physical_device->rad_info.chip_class < GFX7)
return;
diff --git a/src/amd/vulkan/winsys/amdgpu/radv_amdgpu_cs.c b/src/amd/vulkan/winsys/amdgpu/radv_amdgpu_cs.c
index 70f81119c02..0c521917027 100644
--- a/src/amd/vulkan/winsys/amdgpu/radv_amdgpu_cs.c
+++ b/src/amd/vulkan/winsys/amdgpu/radv_amdgpu_cs.c
@@ -1038,5 +1038,5 @@ static int radv_amdgpu_winsys_cs_submit_sysmem(struct radeon_winsys_ctx *_ctx,
bool emit_signal_sem = sem_info->cs_emit_signal;
- if (radv_amdgpu_winsys(ws)->info.chip_class == SI)
+ if (radv_amdgpu_winsys(ws)->info.chip_class == GFX6)
pad_word = 0x80000000;
diff --git a/src/amd/vulkan/winsys/amdgpu/radv_amdgpu_winsys.c b/src/amd/vulkan/winsys/amdgpu/radv_amdgpu_winsys.c
index 35a585a5693..649a7698069 100644
--- a/src/amd/vulkan/winsys/amdgpu/radv_amdgpu_winsys.c
+++ b/src/amd/vulkan/winsys/amdgpu/radv_amdgpu_winsys.c
@@ -59,5 +59,5 @@ do_winsys_init(struct radv_amdgpu_winsys *ws, int fd)
ws->info.num_compute_rings = MIN2(ws->info.num_compute_rings, MAX_RINGS_PER_TYPE);
- ws->use_ib_bos = ws->info.chip_class >= CIK;
+ ws->use_ib_bos = ws->info.chip_class >= GFX7;
return true;
}
diff --git a/src/gallium/drivers/r600/r600_texture.c b/src/gallium/drivers/r600/r600_texture.c
index 27565e0aa0c..497da0c3dfa 100644
--- a/src/gallium/drivers/r600/r600_texture.c
+++ b/src/gallium/drivers/r600/r600_texture.c
@@ -367,5 +367,5 @@ static void r600_reallocate_texture_inplace(struct r600_common_context *rctx,
/* r600g doesn't react to dirty_tex_descriptor_counter */
- if (rctx->chip_class < SI)
+ if (rctx->chip_class < GFX6)
return;
@@ -1265,5 +1265,5 @@ static bool r600_can_invalidate_texture(struct r600_common_screen *rscreen,
{
/* r600g doesn't react to dirty_tex_descriptor_counter */
- return rscreen->chip_class >= SI &&
+ return rscreen->chip_class >= GFX6 &&
!rtex->resource.b.is_shared &&
!(transfer_usage & PIPE_TRANSFER_READ) &&
diff --git a/src/gallium/drivers/r600/r600d_common.h b/src/gallium/drivers/r600/r600d_common.h
index b06f90f8edd..979f26bc7da 100644
--- a/src/gallium/drivers/r600/r600d_common.h
+++ b/src/gallium/drivers/r600/r600d_common.h
@@ -86,5 +86,5 @@
#define SURFACE_BASE_UPDATE_STRMOUT(x) (0x200 << (x))
#define PKT3_SET_SH_REG 0x76 /* SI and later */
-#define PKT3_SET_UCONFIG_REG 0x79 /* CIK and later */
+#define PKT3_SET_UCONFIG_REG 0x79 /* GFX7 and later */
#define EVENT_TYPE_SAMPLE_STREAMOUTSTATS1 0x1 /* EG and later */
diff --git a/src/gallium/drivers/radeonsi/cik_sdma.c b/src/gallium/drivers/radeonsi/cik_sdma.c
index da9b25a442d..2728541dd29 100644
--- a/src/gallium/drivers/radeonsi/cik_sdma.c
+++ b/src/gallium/drivers/radeonsi/cik_sdma.c
@@ -181,10 +181,10 @@ static bool cik_sdma_copy_texture(struct si_context *sctx,
copy_height <= (1 << 14) &&
copy_depth <= (1 << 11) &&
- /* HW limitation - CIK: */
- (sctx->chip_class != CIK ||
+ /* HW limitation - GFX7: */
+ (sctx->chip_class != GFX7 ||
(copy_width < (1 << 14) &&
copy_height < (1 << 14) &&
copy_depth < (1 << 11))) &&
- /* HW limitation - some CIK parts: */
+ /* HW limitation - some GFX7 parts: */
((sctx->family != CHIP_BONAIRE &&
sctx->family != CHIP_KAVERI) ||
@@ -208,5 +208,5 @@ static bool cik_sdma_copy_texture(struct si_context *sctx,
radeon_emit(cs, dstz | ((dst_pitch - 1) << 16));
radeon_emit(cs, dst_slice_pitch - 1);
- if (sctx->chip_class == CIK) {
+ if (sctx->chip_class == GFX7) {
radeon_emit(cs, copy_width | (copy_height << 16));
radeon_emit(cs, copy_depth);
@@ -265,5 +265,5 @@ static bool cik_sdma_copy_texture(struct si_context *sctx,
return false;
- if (sctx->chip_class == CIK &&
+ if (sctx->chip_class == GFX7 &&
(copy_width_aligned == (1 << 14) ||
copy_height == (1 << 14) ||
@@ -372,5 +372,5 @@ static bool cik_sdma_copy_texture(struct si_context *sctx,
radeon_emit(cs, linear_z | ((linear_pitch - 1) << 16));
radeon_emit(cs, linear_slice_pitch - 1);
- if (sctx->chip_class == CIK) {
+ if (sctx->chip_class == GFX7) {
radeon_emit(cs, copy_width_aligned | (copy_height << 16));
radeon_emit(cs, copy_depth);
@@ -395,7 +395,7 @@ static bool cik_sdma_copy_texture(struct si_context *sctx,
srcx % 8 == 0 &&
srcy % 8 == 0 &&
- /* this can either be equal, or display->rotated (VI+ only) */
+ /* this can either be equal, or display->rotated (GFX8+ only) */
(src_micro_mode == dst_micro_mode ||
- (sctx->chip_class >= VI &&
+ (sctx->chip_class >= GFX8 &&
src_micro_mode == V_009910_ADDR_SURF_DISPLAY_MICRO_TILING &&
dst_micro_mode == V_009910_ADDR_SURF_ROTATED_MICRO_TILING))) {
@@ -435,10 +435,10 @@ static bool cik_sdma_copy_texture(struct si_context *sctx,
copy_width_aligned % 8 == 0 &&
copy_height_aligned % 8 == 0 &&
- /* HW limitation - CIK: */
- (sctx->chip_class != CIK ||
+ /* HW limitation - GFX7: */
+ (sctx->chip_class != GFX7 ||
(copy_width_aligned < (1 << 14) &&
copy_height_aligned < (1 << 14) &&
copy_depth < (1 << 11))) &&
- /* HW limitation - some CIK parts: */
+ /* HW limitation - some GFX7 parts: */
((sctx->family != CHIP_BONAIRE &&
sctx->family != CHIP_KAVERI &&
@@ -466,5 +466,5 @@ static bool cik_sdma_copy_texture(struct si_context *sctx,
radeon_emit(cs, dst_slice_tile_max);
radeon_emit(cs, encode_tile_info(sctx, sdst, dst_level, false));
- if (sctx->chip_class == CIK) {
+ if (sctx->chip_class == GFX7) {
radeon_emit(cs, copy_width_aligned |
(copy_height_aligned << 16));
@@ -503,5 +503,5 @@ static void cik_sdma_copy(struct pipe_context *ctx,
}
- if ((sctx->chip_class == CIK || sctx->chip_class == VI) &&
+ if ((sctx->chip_class == GFX7 || sctx->chip_class == GFX8) &&
cik_sdma_copy_texture(sctx, dst, dst_level, dstx, dsty, dstz,
src, src_level, src_box))
diff --git a/src/gallium/drivers/radeonsi/si_blit.c b/src/gallium/drivers/radeonsi/si_blit.c
index 9d3d7d3d27a..5806342cca9 100644
--- a/src/gallium/drivers/radeonsi/si_blit.c
+++ b/src/gallium/drivers/radeonsi/si_blit.c
@@ -1152,5 +1152,5 @@ static bool do_hardware_msaa_resolve(struct pipe_context *ctx,
/* This can happen with mipmapping. */
- if (sctx->chip_class == VI &&
+ if (sctx->chip_class == GFX8 &&
!dst->surface.u.legacy.level[info->dst.level].dcc_fast_clear_size)
goto resolve_to_temp;
diff --git a/src/gallium/drivers/radeonsi/si_clear.c b/src/gallium/drivers/radeonsi/si_clear.c
index d294f236914..d0094031a95 100644
--- a/src/gallium/drivers/radeonsi/si_clear.c
+++ b/src/gallium/drivers/radeonsi/si_clear.c
@@ -314,5 +314,5 @@ static void si_set_optimal_micro_tile_mode(struct si_screen *sscreen,
return;
}
- } else if (sscreen->info.chip_class >= CIK) {
+ } else if (sscreen->info.chip_class >= GFX7) {
/* These magic numbers were copied from addrlib. It doesn't use
* any definitions for them either. They are all 2D_TILED_THIN1
@@ -333,5 +333,5 @@ static void si_set_optimal_micro_tile_mode(struct si_screen *sscreen,
return;
}
- } else { /* SI */
+ } else { /* GFX6 */
switch (tex->last_msaa_resolve_target_micro_mode) {
case RADEON_MICRO_MODE_DISPLAY:
@@ -435,5 +435,5 @@ static void si_do_fast_color_clear(struct si_context *sctx,
continue;
- if (sctx->chip_class <= VI &&
+ if (sctx->chip_class <= GFX8 &&
tex->surface.u.legacy.level[0].mode == RADEON_SURF_MODE_1D &&
!sctx->screen->info.htile_cmask_support_1d_tiling)
@@ -475,5 +475,5 @@ static void si_do_fast_color_clear(struct si_context *sctx,
/* This can happen with mipmapping or MSAA. */
- if (sctx->chip_class == VI &&
+ if (sctx->chip_class == GFX8 &&
!tex->surface.u.legacy.level[level].dcc_fast_clear_size)
continue;
diff --git a/src/gallium/drivers/radeonsi/si_compute.c b/src/gallium/drivers/radeonsi/si_compute.c
index f1a433b72df..46a0ba76ed5 100644
--- a/src/gallium/drivers/radeonsi/si_compute.c
+++ b/src/gallium/drivers/radeonsi/si_compute.c
@@ -328,5 +328,5 @@ static void si_initialize_compute(struct si_context *sctx)
radeon_emit(cs, S_00B85C_SH0_CU_EN(0xffff) | S_00B85C_SH1_CU_EN(0xffff));
- if (sctx->chip_class >= CIK) {
+ if (sctx->chip_class >= GFX7) {
/* Also set R_00B858_COMPUTE_STATIC_THREAD_MGMT_SE2 / SE3 */
radeon_set_sh_reg_seq(cs,
@@ -343,5 +343,5 @@ static void si_initialize_compute(struct si_context *sctx)
* which is now 0x22f.
*/
- if (sctx->chip_class <= SI) {
+ if (sctx->chip_class <= GFX6) {
/* XXX: This should be:
* (number of compute units) * 4 * (waves per simd) - 1 */
@@ -354,5 +354,5 @@ static void si_initialize_compute(struct si_context *sctx)
bc_va = sctx->border_color_buffer->gpu_address;
- if (sctx->chip_class >= CIK) {
+ if (sctx->chip_class >= GFX7) {
radeon_set_uconfig_reg_seq(cs, R_030E00_TA_CS_BC_BASE_ADDR, 2);
radeon_emit(cs, bc_va >> 8); /* R_030E00_TA_CS_BC_BASE_ADDR */
@@ -435,10 +435,10 @@ static bool si_switch_compute_shader(struct si_context *sctx,
lds_blocks = config->lds_size;
- /* XXX: We are over allocating LDS. For SI, the shader reports
+ /* XXX: We are over allocating LDS. For GFX6, the shader reports
* LDS in blocks of 256 bytes, so if there are 4 bytes lds
* allocated in the shader and 4 bytes allocated by the state
* tracker, then we will set LDS_SIZE to 512 bytes rather than 256.
*/
- if (sctx->chip_class <= SI) {
+ if (sctx->chip_class <= GFX6) {
lds_blocks += align(program->local_size, 256) >> 8;
} else {
@@ -475,5 +475,5 @@ static bool si_switch_compute_shader(struct si_context *sctx,
* to get a shader state change in that case anyway.
*/
- if (sctx->chip_class >= CIK) {
+ if (sctx->chip_class >= GFX7) {
cik_prefetch_TC_L2_async(sctx, &program->shader.bo->b.b,
0, program->shader.bo->b.b.width0);
@@ -540,5 +540,5 @@ static void setup_scratch_rsrc_user_sgprs(struct si_context *sctx,
scratch_dword3 |= S_008F0C_ELEMENT_SIZE(max_private_element_size);
- if (sctx->chip_class < VI) {
+ if (sctx->chip_class < GFX8) {
/* BUF_DATA_FORMAT is ignored, but it cannot be
* BUF_DATA_FORMAT_INVALID. */
@@ -765,5 +765,5 @@ static void si_emit_dispatch_packets(struct si_context *sctx,
S_00B854_SIMD_DEST_CNTL(waves_per_threadgroup % 4 == 0);
- if (sctx->chip_class >= CIK) {
+ if (sctx->chip_class >= GFX7) {
unsigned num_cu_per_se = sscreen->info.num_good_compute_units /
sscreen->info.max_se;
@@ -778,5 +778,5 @@ static void si_emit_dispatch_packets(struct si_context *sctx,
compute_resource_limits |= S_00B854_WAVES_PER_SH(sctx->cs_max_waves_per_sh);
} else {
- /* SI */
+ /* GFX6 */
if (sctx->cs_max_waves_per_sh) {
unsigned limit_div16 = DIV_ROUND_UP(sctx->cs_max_waves_per_sh, 16);
@@ -793,5 +793,5 @@ static void si_emit_dispatch_packets(struct si_context *sctx,
/* If the KMD allows it (there is a KMD hw register for it),
* allow launching waves out-of-order. (same as Vulkan) */
- S_00B800_ORDER_MODE(sctx->chip_class >= CIK);
+ S_00B800_ORDER_MODE(sctx->chip_class >= GFX7);
const uint *last_block = info->last_block;
@@ -862,8 +862,8 @@ static void si_launch_grid(
* If async compute is possible, the threadgroup size must be limited
* to 256 threads on all queues to avoid the bug.
- * Only SI and certain CIK chips are affected.
+ * Only GFX6 and certain GFX7 chips are affected.
*/
bool cs_regalloc_hang =
- (sctx->chip_class == SI ||
+ (sctx->chip_class == GFX6 ||
sctx->family == CHIP_BONAIRE ||
sctx->family == CHIP_KABINI) &&
@@ -895,5 +895,5 @@ static void si_launch_grid(
/* Indirect buffers use TC L2 on GFX9, but not older hw. */
- if (sctx->chip_class <= VI &&
+ if (sctx->chip_class <= GFX8 &&
si_resource(info->indirect)->TC_L2_dirty) {
sctx->flags |= SI_CONTEXT_WRITEBACK_GLOBAL_L2;
diff --git a/src/gallium/drivers/radeonsi/si_compute_blit.c b/src/gallium/drivers/radeonsi/si_compute_blit.c
index fb0d8d2f1b6..1cfdc9b62c6 100644
--- a/src/gallium/drivers/radeonsi/si_compute_blit.c
+++ b/src/gallium/drivers/radeonsi/si_compute_blit.c
@@ -37,5 +37,5 @@ static enum si_cache_policy get_cache_policy(struct si_context *sctx,
if ((sctx->chip_class >= GFX9 && (coher == SI_COHERENCY_CB_META ||
coher == SI_COHERENCY_CP)) ||
- (sctx->chip_class >= CIK && coher == SI_COHERENCY_SHADER))
+ (sctx->chip_class >= GFX7 && coher == SI_COHERENCY_SHADER))
return size <= 256 * 1024 ? L2_LRU : L2_STREAM;
@@ -255,5 +255,5 @@ void si_clear_buffer(struct si_context *sctx, struct pipe_resource *dst,
clear_value_size == 4 &&
offset % 4 == 0 &&
- (size > 32*1024 || sctx->chip_class <= VI))) {
+ (size > 32*1024 || sctx->chip_class <= GFX8))) {
si_compute_do_clear_or_copy(sctx, dst, offset, NULL, 0,
aligned_size, clear_value,
@@ -419,5 +419,5 @@ void si_compute_copy_image(struct si_context *sctx,
sctx->flags |= SI_CONTEXT_CS_PARTIAL_FLUSH |
- (sctx->chip_class <= VI ? SI_CONTEXT_WRITEBACK_GLOBAL_L2 : 0) |
+ (sctx->chip_class <= GFX8 ? SI_CONTEXT_WRITEBACK_GLOBAL_L2 : 0) |
si_get_flush_flags(sctx, SI_COHERENCY_SHADER, L2_STREAM);
ctx->bind_compute_state(ctx, saved_cs);
@@ -598,5 +598,5 @@ void si_compute_clear_render_target(struct pipe_context *ctx,
sctx->flags |= SI_CONTEXT_CS_PARTIAL_FLUSH |
- (sctx->chip_class <= VI ? SI_CONTEXT_WRITEBACK_GLOBAL_L2 : 0) |
+ (sctx->chip_class <= GFX8 ? SI_CONTEXT_WRITEBACK_GLOBAL_L2 : 0) |
si_get_flush_flags(sctx, SI_COHERENCY_SHADER, L2_STREAM);
ctx->bind_compute_state(ctx, saved_cs);
diff --git a/src/gallium/drivers/radeonsi/si_cp_dma.c b/src/gallium/drivers/radeonsi/si_cp_dma.c
index 404117d1813..f5c54ca0d52 100644
--- a/src/gallium/drivers/radeonsi/si_cp_dma.c
+++ b/src/gallium/drivers/radeonsi/si_cp_dma.c
@@ -62,5 +62,5 @@ static void si_emit_cp_dma(struct si_context *sctx, struct radeon_cmdbuf *cs,
assert(size <= cp_dma_max_byte_count(sctx));
- assert(sctx->chip_class != SI || cache_policy == L2_BYPASS);
+ assert(sctx->chip_class != GFX6 || cache_policy == L2_BYPASS);
if (sctx->chip_class >= GFX9)
@@ -91,5 +91,5 @@ static void si_emit_cp_dma(struct si_context *sctx, struct radeon_cmdbuf *cs,
command |= S_414_DAS(V_414_REGISTER) |
S_414_DAIC(V_414_NO_INCREMENT);
- } else if (sctx->chip_class >= CIK && cache_policy != L2_BYPASS) {
+ } else if (sctx->chip_class >= GFX7 && cache_policy != L2_BYPASS) {
header |= S_411_DST_SEL(V_411_DST_ADDR_TC_L2) |
S_500_DST_CACHE_POLICY(cache_policy == L2_STREAM);
@@ -103,10 +103,10 @@ static void si_emit_cp_dma(struct si_context *sctx, struct radeon_cmdbuf *cs,
command |= S_414_SAS(V_414_REGISTER) |
S_414_SAIC(V_414_NO_INCREMENT);
- } else if (sctx->chip_class >= CIK && cache_policy != L2_BYPASS) {
+ } else if (sctx->chip_class >= GFX7 && cache_policy != L2_BYPASS) {
header |= S_411_SRC_SEL(V_411_SRC_ADDR_TC_L2) |
S_500_SRC_CACHE_POLICY(cache_policy == L2_STREAM);
}
- if (sctx->chip_class >= CIK) {
+ if (sctx->chip_class >= GFX7) {
radeon_emit(cs, PKT3(PKT3_DMA_DATA, 5, 0));
radeon_emit(cs, header);
@@ -413,5 +413,5 @@ void cik_prefetch_TC_L2_async(struct si_context *sctx, struct pipe_resource *buf
uint64_t offset, unsigned size)
{
- assert(sctx->chip_class >= CIK);
+ assert(sctx->chip_class >= GFX7);
si_cp_dma_copy_buffer(sctx, buf, buf, offset, offset, size,
@@ -492,5 +492,5 @@ void cik_emit_prefetch_L2(struct si_context *sctx, bool vertex_stage_only)
}
} else {
- /* SI-CI-VI */
+ /* GFX6-GFX8 */
/* Choose the right spot for the VBO prefetch. */
if (sctx->tes_shader.cso) {
@@ -592,5 +592,5 @@ void si_cp_write_data(struct si_context *sctx, struct si_resource *buf,
assert(size % 4 == 0);
- if (sctx->chip_class == SI && dst_sel == V_370_MEM)
+ if (sctx->chip_class == GFX6 && dst_sel == V_370_MEM)
dst_sel = V_370_MEM_GRBM;
diff --git a/src/gallium/drivers/radeonsi/si_debug.c b/src/gallium/drivers/radeonsi/si_debug.c
index 9a4494a98fe..bd85fc49387 100644
--- a/src/gallium/drivers/radeonsi/si_debug.c
+++ b/src/gallium/drivers/radeonsi/si_debug.c
@@ -315,5 +315,5 @@ static void si_dump_debug_registers(struct si_context *sctx, FILE *f)
si_dump_mmapped_reg(sctx, f, R_00D034_SDMA0_STATUS_REG);
si_dump_mmapped_reg(sctx, f, R_00D834_SDMA1_STATUS_REG);
- if (sctx->chip_class <= VI) {
+ if (sctx->chip_class <= GFX8) {
si_dump_mmapped_reg(sctx, f, R_000E50_SRBM_STATUS);
si_dump_mmapped_reg(sctx, f, R_000E4C_SRBM_STATUS2);
diff --git a/src/gallium/drivers/radeonsi/si_descriptors.c b/src/gallium/drivers/radeonsi/si_descriptors.c
index 68b5430446e..7234e4767a3 100644
--- a/src/gallium/drivers/radeonsi/si_descriptors.c
+++ b/src/gallium/drivers/radeonsi/si_descriptors.c
@@ -348,5 +348,5 @@ void si_set_mutable_tex_desc_fields(struct si_screen *sscreen,
state[0] |= tex->surface.tile_swizzle;
- if (sscreen->info.chip_class >= VI) {
+ if (sscreen->info.chip_class >= GFX8) {
state[6] &= C_008F28_COMPRESSION_EN;
state[7] = 0;
@@ -356,5 +356,5 @@ void si_set_mutable_tex_desc_fields(struct si_screen *sscreen,
tex->dcc_offset;
- if (sscreen->info.chip_class == VI) {
+ if (sscreen->info.chip_class == GFX8) {
meta_va += base_level_info->dcc_offset;
assert(base_level_info->mode == RADEON_SURF_MODE_2D);
@@ -400,5 +400,5 @@ void si_set_mutable_tex_desc_fields(struct si_screen *sscreen,
}
} else {
- /* SI-CI-VI */
+ /* GFX6-GFX8 */
unsigned pitch = base_level_info->nblk_x * block_width;
unsigned index = si_tile_mode_index(tex, base_level, is_stencil);
@@ -1144,5 +1144,5 @@ bool si_upload_vertex_buffer_descriptors(struct si_context *sctx)
int64_t num_records = (int64_t)buf->b.b.width0 - offset;
- if (sctx->chip_class != VI && vb->stride) {
+ if (sctx->chip_class != GFX8 && vb->stride) {
/* Round up by rounding down and adding 1 */
num_records = (num_records - velems->format_size[i]) /
@@ -1213,7 +1213,7 @@ static void si_set_constant_buffer(struct si_context *sctx,
pipe_resource_reference(&buffers->buffers[slot], NULL);
- /* CIK cannot unbind a constant buffer (S_BUFFER_LOAD is buggy
+ /* GFX7 cannot unbind a constant buffer (S_BUFFER_LOAD is buggy
* with a NULL buffer). We need to use a dummy buffer instead. */
- if (sctx->chip_class == CIK &&
+ if (sctx->chip_class == GFX7 &&
(!input || (!input->buffer && !input->user_buffer)))
input = &sctx->null_const_buf;
@@ -1472,5 +1472,5 @@ void si_set_ring_buffer(struct si_context *sctx, uint slot,
}
- if (sctx->chip_class >= VI && stride)
+ if (sctx->chip_class >= GFX8 && stride)
num_records *= stride;
diff --git a/src/gallium/drivers/radeonsi/si_dma_cs.c b/src/gallium/drivers/radeonsi/si_dma_cs.c
index bba1bd95826..8f2e15833b6 100644
--- a/src/gallium/drivers/radeonsi/si_dma_cs.c
+++ b/src/gallium/drivers/radeonsi/si_dma_cs.c
@@ -31,5 +31,5 @@ static void si_dma_emit_wait_idle(struct si_context *sctx)
/* NOP waits for idle. */
- if (sctx->chip_class >= CIK)
+ if (sctx->chip_class >= GFX7)
radeon_emit(cs, 0x00000000); /* NOP */
else
@@ -43,5 +43,5 @@ void si_dma_emit_timestamp(struct si_context *sctx, struct si_resource *dst,
uint64_t va = dst->gpu_address + offset;
- if (sctx->chip_class == SI) {
+ if (sctx->chip_class == GFX6) {
unreachable("SI DMA doesn't support the timestamp packet.");
return;
@@ -88,5 +88,5 @@ void si_sdma_clear_buffer(struct si_context *sctx, struct pipe_resource *dst,
offset += sdst->gpu_address;
- if (sctx->chip_class == SI) {
+ if (sctx->chip_class == GFX6) {
/* the same maximum size as for copying */
ncopy = DIV_ROUND_UP(size, SI_DMA_COPY_MAX_DWORD_ALIGNED_SIZE);
@@ -106,5 +106,5 @@ void si_sdma_clear_buffer(struct si_context *sctx, struct pipe_resource *dst,
}
- /* The following code is for CI, VI, Vega/Raven, etc. */
+ /* The following code is for Sea Islands and later. */
/* the same maximum size as for copying */
ncopy = DIV_ROUND_UP(size, CIK_SDMA_COPY_MAX_SIZE);
diff --git a/src/gallium/drivers/radeonsi/si_fence.c b/src/gallium/drivers/radeonsi/si_fence.c
index ffda98d2834..b3212c1db35 100644
--- a/src/gallium/drivers/radeonsi/si_fence.c
+++ b/src/gallium/drivers/radeonsi/si_fence.c
@@ -116,6 +116,6 @@ void si_cp_release_mem(struct si_context *ctx,
radeon_emit(cs, 0); /* unused */
} else {
- if (ctx->chip_class == CIK ||
- ctx->chip_class == VI) {
+ if (ctx->chip_class == GFX7 ||
+ ctx->chip_class == GFX8) {
struct si_resource *scratch = ctx->eop_bug_scratch;
uint64_t va = scratch->gpu_address;
@@ -154,6 +154,6 @@ unsigned si_cp_write_fence_dwords(struct si_screen *screen)
unsigned dwords = 6;
- if (screen->info.chip_class == CIK ||
- screen->info.chip_class == VI)
+ if (screen->info.chip_class == GFX7 ||
+ screen->info.chip_class == GFX8)
dwords *= 2;
diff --git a/src/gallium/drivers/radeonsi/si_get.c b/src/gallium/drivers/radeonsi/si_get.c
index 5593d3bac05..c2dc25b21db 100644
--- a/src/gallium/drivers/radeonsi/si_get.c
+++ b/src/gallium/drivers/radeonsi/si_get.c
@@ -254,5 +254,5 @@ static int si_get_param(struct pipe_screen *pscreen, enum pipe_cap param)
case PIPE_CAP_TEXTURE_BORDER_COLOR_QUIRK:
- return sscreen->info.chip_class <= VI ?
+ return sscreen->info.chip_class <= GFX8 ?
PIPE_QUIRK_TEXTURE_BORDER_COLOR_SWIZZLE_R600 : 0;
diff --git a/src/gallium/drivers/radeonsi/si_gfx_cs.c b/src/gallium/drivers/radeonsi/si_gfx_cs.c
index d0d405c473f..c81718950a4 100644
--- a/src/gallium/drivers/radeonsi/si_gfx_cs.c
+++ b/src/gallium/drivers/radeonsi/si_gfx_cs.c
@@ -83,5 +83,5 @@ void si_flush_gfx_cs(struct si_context *ctx, unsigned flags,
SI_CONTEXT_CS_PARTIAL_FLUSH |
SI_CONTEXT_INV_GLOBAL_L2;
- } else if (ctx->chip_class == SI) {
+ } else if (ctx->chip_class == GFX6) {
/* The kernel flushes L2 before shaders are finished. */
wait_flags |= SI_CONTEXT_PS_PARTIAL_FLUSH |
@@ -148,5 +148,5 @@ void si_flush_gfx_cs(struct si_context *ctx, unsigned flags,
/* Make sure CP DMA is idle at the end of IBs after L2 prefetches
* because the kernel doesn't wait for it. */
- if (ctx->chip_class >= CIK)
+ if (ctx->chip_class >= GFX7)
si_cp_dma_wait_for_idle(ctx);
@@ -408,5 +408,5 @@ void si_begin_new_gfx_cs(struct si_context *ctx)
ctx->tracked_regs.reg_value[SI_TRACKED_CB_SHADER_MASK] = 0xffffffff;
ctx->tracked_regs.reg_value[SI_TRACKED_VGT_TF_PARAM] = 0x00000000;
- ctx->tracked_regs.reg_value[SI_TRACKED_VGT_VERTEX_REUSE_BLOCK_CNTL] = 0x0000001e; /* From VI */
+ ctx->tracked_regs.reg_value[SI_TRACKED_VGT_VERTEX_REUSE_BLOCK_CNTL] = 0x0000001e; /* From GFX8 */
/* Set all saved registers state to saved. */
diff --git a/src/gallium/drivers/radeonsi/si_gpu_load.c b/src/gallium/drivers/radeonsi/si_gpu_load.c
index 481438f37bb..7c2e43b3fdd 100644
--- a/src/gallium/drivers/radeonsi/si_gpu_load.c
+++ b/src/gallium/drivers/radeonsi/si_gpu_load.c
@@ -103,5 +103,5 @@ static void si_update_mmio_counters(struct si_screen *sscreen,
gui_busy = GUI_ACTIVE(value);
- if (sscreen->info.chip_class == CIK || sscreen->info.chip_class == VI) {
+ if (sscreen->info.chip_class == GFX7 || sscreen->info.chip_class == GFX8) {
/* SRBM_STATUS2 */
sscreen->ws->read_registers(sscreen->ws, SRBM_STATUS2, 1, &value);
@@ -111,5 +111,5 @@ static void si_update_mmio_counters(struct si_screen *sscreen,
}
- if (sscreen->info.chip_class >= VI) {
+ if (sscreen->info.chip_class >= GFX8) {
/* CP_STAT */
sscreen->ws->read_registers(sscreen->ws, CP_STAT, 1, &value);
diff --git a/src/gallium/drivers/radeonsi/si_perfcounter.c b/src/gallium/drivers/radeonsi/si_perfcounter.c
index c15c444cc40..322950557e3 100644
--- a/src/gallium/drivers/radeonsi/si_perfcounter.c
+++ b/src/gallium/drivers/radeonsi/si_perfcounter.c
@@ -1285,9 +1285,9 @@ void si_init_perfcounters(struct si_screen *screen)
switch (screen->info.chip_class) {
- case CIK:
+ case GFX7:
blocks = groups_CIK;
num_blocks = ARRAY_SIZE(groups_CIK);
break;
- case VI:
+ case GFX8:
blocks = groups_VI;
num_blocks = ARRAY_SIZE(groups_VI);
@@ -1297,5 +1297,5 @@ void si_init_perfcounters(struct si_screen *screen)
num_blocks = ARRAY_SIZE(groups_gfx9);
break;
- case SI:
+ case GFX6:
default:
return; /* not implemented */
@@ -1303,5 +1303,5 @@ void si_init_perfcounters(struct si_screen *screen)
if (screen->info.max_sh_per_se != 1) {
- /* This should not happen on non-SI chips. */
+ /* This should not happen on non-GFX6 chips. */
fprintf(stderr, "si_init_perfcounters: max_sh_per_se = %d not "
"supported (inaccurate performance counters)\n",
diff --git a/src/gallium/drivers/radeonsi/si_pipe.c b/src/gallium/drivers/radeonsi/si_pipe.c
index 95280675506..0c1324332d9 100644
--- a/src/gallium/drivers/radeonsi/si_pipe.c
+++ b/src/gallium/drivers/radeonsi/si_pipe.c
@@ -116,5 +116,5 @@ static void si_init_compiler(struct si_screen *sscreen,
* predating Ryzen (Raven). */
bool create_low_opt_compiler = !sscreen->info.has_dedicated_vram &&
- sscreen->info.chip_class <= VI;
+ sscreen->info.chip_class <= GFX8;
enum ac_target_machine_options tm_options =
@@ -378,5 +378,5 @@ static struct pipe_context *si_create_context(struct pipe_screen *screen,
return NULL;
- sctx->has_graphics = sscreen->info.chip_class == SI ||
+ sctx->has_graphics = sscreen->info.chip_class == GFX6 ||
!(flags & PIPE_CONTEXT_COMPUTE_ONLY);
@@ -397,6 +397,6 @@ static struct pipe_context *si_create_context(struct pipe_screen *screen,
sctx->chip_class = sscreen->info.chip_class;
- if (sctx->chip_class == CIK ||
- sctx->chip_class == VI ||
+ if (sctx->chip_class == GFX7 ||
+ sctx->chip_class == GFX8 ||
sctx->chip_class == GFX9) {
sctx->eop_bug_scratch = si_resource(
@@ -514,5 +514,5 @@ static struct pipe_context *si_create_context(struct pipe_screen *screen,
/* Initialize SDMA functions. */
- if (sctx->chip_class >= CIK)
+ if (sctx->chip_class >= GFX7)
cik_init_sdma_functions(sctx);
else
@@ -541,7 +541,7 @@ static struct pipe_context *si_create_context(struct pipe_screen *screen,
}
- /* CIK cannot unbind a constant buffer (S_BUFFER_LOAD doesn't skip loads
+ /* GFX7 cannot unbind a constant buffer (S_BUFFER_LOAD doesn't skip loads
* if NUM_RECORDS == 0). We need to use a dummy buffer instead. */
- if (sctx->chip_class == CIK) {
+ if (sctx->chip_class == GFX7) {
sctx->null_const_buf.buffer =
pipe_aligned_buffer_create(screen,
@@ -616,5 +616,5 @@ static struct pipe_context *si_create_context(struct pipe_screen *screen,
si_begin_new_gfx_cs(sctx);
- if (sctx->chip_class == CIK) {
+ if (sctx->chip_class == GFX7) {
/* Clear the NULL constant buffer, because loads should return zeros.
* Note that this forces CP DMA to be used, because clover deadlocks
@@ -995,9 +995,9 @@ struct pipe_screen *radeonsi_screen_create(struct radeon_winsys *ws,
/* Determine tessellation ring info. */
- bool double_offchip_buffers = sscreen->info.chip_class >= CIK &&
+ bool double_offchip_buffers = sscreen->info.chip_class >= GFX7 &&
sscreen->info.family != CHIP_CARRIZO &&
sscreen->info.family != CHIP_STONEY;
/* This must be one less than the maximum number due to a hw limitation.
- * Various hardware bugs in SI, CIK, and GFX9 need this.
+ * Various hardware bugs need this.
*/
unsigned max_offchip_buffers_per_se;
@@ -1030,6 +1030,6 @@ struct pipe_screen *radeonsi_screen_create(struct radeon_winsys *ws,
sscreen->tess_offchip_block_dw_size * 4;
- if (sscreen->info.chip_class >= CIK) {
- if (sscreen->info.chip_class >= VI)
+ if (sscreen->info.chip_class >= GFX7) {
+ if (sscreen->info.chip_class >= GFX8)
--max_offchip_buffers;
sscreen->vgt_hs_offchip_param =
@@ -1043,26 +1043,26 @@ struct pipe_screen *radeonsi_screen_create(struct radeon_winsys *ws,
/* The mere presense of CLEAR_STATE in the IB causes random GPU hangs
- * on SI. Some CLEAR_STATE cause asic hang on radeon kernel, etc.
- * SPI_VS_OUT_CONFIG. So only enable CI CLEAR_STATE on amdgpu kernel.*/
- sscreen->has_clear_state = sscreen->info.chip_class >= CIK &&
+ * on GFX6. Some CLEAR_STATE cause asic hang on radeon kernel, etc.
+ * SPI_VS_OUT_CONFIG. So only enable GFX7 CLEAR_STATE on amdgpu kernel.*/
+ sscreen->has_clear_state = sscreen->info.chip_class >= GFX7 &&
sscreen->info.drm_major == 3;
sscreen->has_distributed_tess =
- sscreen->info.chip_class >= VI &&
+ sscreen->info.chip_class >= GFX8 &&
sscreen->info.max_se >= 2;
sscreen->has_draw_indirect_multi =
(sscreen->info.family >= CHIP_POLARIS10) ||
- (sscreen->info.chip_class == VI &&
+ (sscreen->info.chip_class == GFX8 &&
sscreen->info.pfp_fw_version >= 121 &&
sscreen->info.me_fw_version >= 87) ||
- (sscreen->info.chip_class == CIK &&
+ (sscreen->info.chip_class == GFX7 &&
sscreen->info.pfp_fw_version >= 211 &&
sscreen->info.me_fw_version >= 173) ||
- (sscreen->info.chip_class == SI &&
+ (sscreen->info.chip_class == GFX6 &&
sscreen->info.pfp_fw_version >= 79 &&
sscreen->info.me_fw_version >= 142);
- sscreen->has_out_of_order_rast = sscreen->info.chip_class >= VI &&
+ sscreen->has_out_of_order_rast = sscreen->info.chip_class >= GFX8 &&
sscreen->info.max_se >= 2 &&
!(sscreen->debug_flags & DBG(NO_OUT_OF_ORDER));
@@ -1115,5 +1115,5 @@ struct pipe_screen *radeonsi_screen_create(struct radeon_winsys *ws,
* on GFX9.
*/
- sscreen->llvm_has_working_vgpr_indexing = sscreen->info.chip_class <= VI;
+ sscreen->llvm_has_working_vgpr_indexing = sscreen->info.chip_class <= GFX8;
/* Some chips have RB+ registers, but don't support RB+. Those must
@@ -1135,5 +1135,5 @@ struct pipe_screen *radeonsi_screen_create(struct radeon_winsys *ws,
!(sscreen->debug_flags & DBG(NO_DCC_MSAA));
- sscreen->cpdma_prefetch_writes_memory = sscreen->info.chip_class <= VI;
+ sscreen->cpdma_prefetch_writes_memory = sscreen->info.chip_class <= GFX8;
(void) mtx_init(&sscreen->shader_parts_mutex, mtx_plain);
@@ -1143,5 +1143,5 @@ struct pipe_screen *radeonsi_screen_create(struct radeon_winsys *ws,
sscreen->barrier_flags.cp_to_L2 = SI_CONTEXT_INV_SMEM_L1 |
SI_CONTEXT_INV_VMEM_L1;
- if (sscreen->info.chip_class <= VI) {
+ if (sscreen->info.chip_class <= GFX8) {
sscreen->barrier_flags.cp_to_L2 |= SI_CONTEXT_INV_GLOBAL_L2;
sscreen->barrier_flags.L2_to_cp |= SI_CONTEXT_WRITEBACK_GLOBAL_L2;
diff --git a/src/gallium/drivers/radeonsi/si_pipe.h b/src/gallium/drivers/radeonsi/si_pipe.h
index 32f0efd1713..d3792b361de 100644
--- a/src/gallium/drivers/radeonsi/si_pipe.h
+++ b/src/gallium/drivers/radeonsi/si_pipe.h
@@ -73,5 +73,5 @@
#define SI_CONTEXT_INV_GLOBAL_L2 (1 << 6)
/* Write dirty L2 lines back to memory (shader and CP DMA stores), but don't
- * invalidate L2. SI-CIK can't do it, so they will do complete invalidation. */
+ * invalidate L2. GFX6-GFX7 can't do it, so they will do complete invalidation. */
#define SI_CONTEXT_WRITEBACK_GLOBAL_L2 (1 << 7)
/* Writeback & invalidate the L2 metadata cache. It can only be coupled with
@@ -367,5 +367,5 @@ struct si_surface {
unsigned cb_color_attrib;
unsigned cb_color_attrib2; /* GFX9 and later */
- unsigned cb_dcc_control; /* VI and later */
+ unsigned cb_dcc_control; /* GFX8 and later */
unsigned spi_shader_col_format:8; /* no blending, no alpha-to-coverage. */
unsigned spi_shader_col_format_alpha:8; /* alpha-to-coverage */
@@ -925,5 +925,5 @@ struct si_context {
/* other shader resources */
- struct pipe_constant_buffer null_const_buf; /* used for set_constant_buffer(NULL) on CIK */
+ struct pipe_constant_buffer null_const_buf; /* used for set_constant_buffer(NULL) on GFX7 */
struct pipe_resource *esgs_ring;
struct pipe_resource *gsvs_ring;
@@ -1574,5 +1574,5 @@ si_make_CB_shader_coherent(struct si_context *sctx, unsigned num_samples,
sctx->flags |= SI_CONTEXT_INV_L2_METADATA;
} else {
- /* SI-CI-VI */
+ /* GFX6-GFX8 */
sctx->flags |= SI_CONTEXT_INV_GLOBAL_L2;
}
@@ -1596,5 +1596,5 @@ si_make_DB_shader_coherent(struct si_context *sctx, unsigned num_samples,
sctx->flags |= SI_CONTEXT_INV_L2_METADATA;
} else {
- /* SI-CI-VI */
+ /* GFX6-GFX8 */
sctx->flags |= SI_CONTEXT_INV_GLOBAL_L2;
}
diff --git a/src/gallium/drivers/radeonsi/si_pm4.c b/src/gallium/drivers/radeonsi/si_pm4.c
index 22c4a5b6e6e..0b7d53e745d 100644
--- a/src/gallium/drivers/radeonsi/si_pm4.c
+++ b/src/gallium/drivers/radeonsi/si_pm4.c
@@ -162,6 +162,6 @@ void si_pm4_upload_indirect_buffer(struct si_context *sctx,
unsigned aligned_ndw = align(state->ndw, 8);
- /* only supported on CIK and later */
- if (sctx->chip_class < CIK)
+ /* only supported on GFX7 and later */
+ if (sctx->chip_class < GFX7)
return;
diff --git a/src/gallium/drivers/radeonsi/si_query.c b/src/gallium/drivers/radeonsi/si_query.c
index 3e357e8b6c0..c233141f0c1 100644
--- a/src/gallium/drivers/radeonsi/si_query.c
+++ b/src/gallium/drivers/radeonsi/si_query.c
@@ -1020,5 +1020,5 @@ static void si_emit_query_predication(struct si_context *ctx)
* the wait flag does not apply in this predication mode.
*
- * The shader outputs the result value to L2. Workarounds only affect VI
+ * The shader outputs the result value to L2. Workarounds only affect GFX8
* and later, where the CP reads data from L2, so we don't need an
* additional flush.
@@ -1609,9 +1609,9 @@ static void si_render_condition(struct pipe_context *ctx,
bool needs_workaround = false;
- /* There was a firmware regression in VI which causes successive
+ /* There was a firmware regression in GFX8 which causes successive
* SET_PREDICATION packets to give the wrong answer for
* non-inverted stream overflow predication.
*/
- if (((sctx->chip_class == VI && sctx->screen->info.pfp_fw_feature < 49) ||
+ if (((sctx->chip_class == GFX8 && sctx->screen->info.pfp_fw_feature < 49) ||
(sctx->chip_class == GFX9 && sctx->screen->info.pfp_fw_feature < 38)) &&
!condition &&
@@ -1787,5 +1787,5 @@ static unsigned si_get_num_queries(struct si_screen *sscreen)
/* amdgpu */
if (sscreen->info.drm_major == 3) {
- if (sscreen->info.chip_class >= VI)
+ if (sscreen->info.chip_class >= GFX8)
return ARRAY_SIZE(si_driver_query_list);
else
@@ -1795,5 +1795,5 @@ static unsigned si_get_num_queries(struct si_screen *sscreen)
/* radeon */
if (sscreen->info.has_read_registers_query) {
- if (sscreen->info.chip_class == CIK)
+ if (sscreen->info.chip_class == GFX7)
return ARRAY_SIZE(si_driver_query_list) - 6;
else
diff --git a/src/gallium/drivers/radeonsi/si_shader.c b/src/gallium/drivers/radeonsi/si_shader.c
index f6d882cf583..98c11e1c98d 100644
--- a/src/gallium/drivers/radeonsi/si_shader.c
+++ b/src/gallium/drivers/radeonsi/si_shader.c
@@ -106,5 +106,5 @@ static bool llvm_type_is_64bit(struct si_shader_context *ctx,
static bool is_merged_shader(struct si_shader_context *ctx)
{
- if (ctx->screen->info.chip_class <= VI)
+ if (ctx->screen->info.chip_class <= GFX8)
return false;
@@ -3083,5 +3083,5 @@ static void si_write_tess_factors(struct lp_build_tgsi_context *bld_base,
/* Store the dynamic HS control word. */
offset = 0;
- if (ctx->screen->info.chip_class <= VI) {
+ if (ctx->screen->info.chip_class <= GFX8) {
ac_build_buffer_store_dword(&ctx->ac, buffer,
LLVMConstInt(ctx->i32, 0x80000000, 0),
@@ -3720,5 +3720,5 @@ static void si_llvm_return_fs_outputs(struct ac_shader_abi *abi,
break;
default:
- fprintf(stderr, "Warning: SI unhandled fs output type:%d\n",
+ fprintf(stderr, "Warning: GFX6 unhandled fs output type:%d\n",
semantic_name);
}
@@ -4216,9 +4216,9 @@ static void si_llvm_emit_barrier(const struct lp_build_tgsi_action *action,
struct si_shader_context *ctx = si_shader_context(bld_base);
- /* SI only (thanks to a hw bug workaround):
+ /* GFX6 only (thanks to a hw bug workaround):
* The real barrier instruction isn’t needed, because an entire patch
* always fits into a single wave.
*/
- if (ctx->screen->info.chip_class == SI &&
+ if (ctx->screen->info.chip_class == GFX6 &&
ctx->type == PIPE_SHADER_TESS_CTRL) {
ac_build_waitcnt(&ctx->ac, LGKM_CNT & VM_CNT);
@@ -4328,5 +4328,5 @@ static unsigned si_get_max_workgroup_size(const struct si_shader *shader)
/* Return this so that LLVM doesn't remove s_barrier
* instructions on chips where we use s_barrier. */
- return shader->selector->screen->info.chip_class >= CIK ? 128 : 64;
+ return shader->selector->screen->info.chip_class >= GFX7 ? 128 : 64;
case PIPE_SHADER_GEOMETRY:
@@ -4537,5 +4537,5 @@ static void create_function(struct si_shader_context *ctx)
break;
- case PIPE_SHADER_TESS_CTRL: /* SI-CI-VI */
+ case PIPE_SHADER_TESS_CTRL: /* GFX6-GFX8 */
declare_global_desc_pointers(ctx, &fninfo);
declare_per_stage_desc_pointers(ctx, &fninfo, true);
@@ -4852,5 +4852,5 @@ static void preload_ring_buffers(struct si_shader_context *ctx)
ctx->param_rw_buffers);
- if (ctx->screen->info.chip_class <= VI &&
+ if (ctx->screen->info.chip_class <= GFX8 &&
(ctx->shader->key.as_es || ctx->type == PIPE_SHADER_GEOMETRY)) {
unsigned ring =
@@ -4898,5 +4898,5 @@ static void preload_ring_buffers(struct si_shader_context *ctx)
stride = 4 * num_components * sel->gs_max_out_vertices;
- /* Limit on the stride field for <= CIK. */
+ /* Limit on the stride field for <= GFX7. */
assert(stride < (1 << 14));
@@ -5223,5 +5223,5 @@ static void si_calculate_max_simd_waves(struct si_shader *shader)
struct si_shader_config *conf = &shader->config;
unsigned num_inputs = shader->selector->info.num_inputs;
- unsigned lds_increment = sscreen->info.chip_class >= CIK ? 512 : 256;
+ unsigned lds_increment = sscreen->info.chip_class >= GFX7 ? 512 : 256;
unsigned lds_per_wave = 0;
unsigned max_simd_waves;
@@ -5453,5 +5453,5 @@ static int si_compile_llvm(struct si_screen *sscreen,
* - Some opcodes don't support denormals, such as v_mad_f32. We would
* have to stop using those.
- * - SI & CI would be very slow.
+ * - GFX6 & GFX7 would be very slow.
*/
conf->float_mode |= V_00B028_FP_64_DENORMS;
@@ -6577,5 +6577,5 @@ static bool si_should_optimize_less(struct ac_llvm_compiler *compiler,
/* Assume a slow CPU. */
assert(!sel->screen->info.has_dedicated_vram &&
- sel->screen->info.chip_class <= VI);
+ sel->screen->info.chip_class <= GFX8);
/* For a crazy dEQP test containing 2597 memory opcodes, mostly
@@ -6832,5 +6832,5 @@ int si_compile_tgsi_shader(struct si_screen *sscreen,
unsigned wave_size = 64;
unsigned max_vgprs = 256;
- unsigned max_sgprs = sscreen->info.chip_class >= VI ? 800 : 512;
+ unsigned max_sgprs = sscreen->info.chip_class >= GFX8 ? 800 : 512;
unsigned max_sgprs_per_wave = 128;
unsigned max_block_threads = si_get_max_workgroup_size(shader);
@@ -7264,5 +7264,5 @@ static void si_build_tcs_epilog_function(struct si_shader_context *ctx,
/* Create the function. */
si_create_function(ctx, "tcs_epilog", NULL, 0, &fninfo,
- ctx->screen->info.chip_class >= CIK ? 128 : 64);
+ ctx->screen->info.chip_class >= GFX7 ? 128 : 64);
ac_declare_lds_as_pointer(&ctx->ac);
func = ctx->main_fn;
diff --git a/src/gallium/drivers/radeonsi/si_shader.h b/src/gallium/drivers/radeonsi/si_shader.h
index 82c521efcb7..16b78fbf43e 100644
--- a/src/gallium/drivers/radeonsi/si_shader.h
+++ b/src/gallium/drivers/radeonsi/si_shader.h
@@ -248,5 +248,5 @@ enum {
#define C_VS_STATE_LS_OUT_VERTEX_SIZE 0x00FFFFFF
-/* SI-specific system values. */
+/* Driver-specific system values. */
enum {
/* Values from set_tess_state. */
diff --git a/src/gallium/drivers/radeonsi/si_shader_tgsi_mem.c b/src/gallium/drivers/radeonsi/si_shader_tgsi_mem.c
index 5e540fc5098..be0cb89f722 100644
--- a/src/gallium/drivers/radeonsi/si_shader_tgsi_mem.c
+++ b/src/gallium/drivers/radeonsi/si_shader_tgsi_mem.c
@@ -49,6 +49,6 @@ static LLVMValueRef get_buffer_size(
LLVMConstInt(ctx->i32, 2, 0), "");
- if (ctx->screen->info.chip_class == VI) {
- /* On VI, the descriptor contains the size in bytes,
+ if (ctx->screen->info.chip_class == GFX8) {
+ /* On GFX8, the descriptor contains the size in bytes,
* but TXQ must return the size in elements.
* The stride is always non-zero for resources using TXQ.
@@ -133,5 +133,5 @@ ac_image_dim_from_tgsi_target(struct si_screen *screen, enum tgsi_texture_type t
/* Match the resource type set in the descriptor. */
if (dim == ac_image_cube ||
- (screen->info.chip_class <= VI && dim == ac_image_3d))
+ (screen->info.chip_class <= GFX8 && dim == ac_image_3d))
dim = ac_image_2darray;
else if (target == TGSI_TEXTURE_2D && screen->info.chip_class >= GFX9) {
@@ -162,5 +162,5 @@ static LLVMValueRef force_dcc_off(struct si_shader_context *ctx,
LLVMValueRef rsrc)
{
- if (ctx->screen->info.chip_class <= CIK) {
+ if (ctx->screen->info.chip_class <= GFX7) {
return rsrc;
} else {
@@ -328,9 +328,9 @@ static unsigned get_cache_policy(struct si_shader_context *ctx,
if (!atomic &&
- /* SI has a TC L1 bug causing corruption of 8bit/16bit stores.
+ /* GFX6 has a TC L1 bug causing corruption of 8bit/16bit stores.
* All store opcodes not aligned to a dword are affected.
* The only way to get unaligned stores in radeonsi is through
* shader images. */
- ((may_store_unaligned && ctx->screen->info.chip_class == SI) ||
+ ((may_store_unaligned && ctx->screen->info.chip_class == GFX6) ||
/* If this is write-only, don't keep data in L1 to prevent
* evicting L1 cache lines that may be needed by other
@@ -1100,5 +1100,5 @@ LLVMValueRef si_load_sampler_desc(struct si_shader_context *ctx,
/* Disable anisotropic filtering if BASE_LEVEL == LAST_LEVEL.
*
- * SI-CI:
+ * GFX6-GFX7:
* If BASE_LEVEL == LAST_LEVEL, the shader must disable anisotropic
* filtering manually. The driver sets img7 to a mask clearing
@@ -1106,5 +1106,5 @@ LLVMValueRef si_load_sampler_desc(struct si_shader_context *ctx,
* s_and_b32 samp0, samp0, img7
*
- * VI:
+ * GFX8:
* The ANISO_OVERRIDE sampler field enables this fix in TA.
*/
@@ -1114,5 +1114,5 @@ static LLVMValueRef sici_fix_sampler_aniso(struct si_shader_context *ctx,
LLVMValueRef img7, samp0;
- if (ctx->screen->info.chip_class >= VI)
+ if (ctx->screen->info.chip_class >= GFX8)
return samp;
@@ -1447,5 +1447,5 @@ static void build_tex_intrinsic(const struct lp_build_tgsi_action *action,
* Z24 anymore. Do it manually here.
*/
- if (ctx->screen->info.chip_class >= VI) {
+ if (ctx->screen->info.chip_class >= GFX8) {
LLVMValueRef upgraded;
LLVMValueRef clamped;
@@ -1531,5 +1531,5 @@ static void build_tex_intrinsic(const struct lp_build_tgsi_action *action,
opcode != TGSI_OPCODE_TXF &&
opcode != TGSI_OPCODE_TXF_LZ &&
- ctx->screen->info.chip_class <= VI) {
+ ctx->screen->info.chip_class <= GFX8) {
unsigned array_coord = target == TGSI_TEXTURE_1D_ARRAY ? 1 : 2;
args.coords[array_coord] = ac_build_round(&ctx->ac, args.coords[array_coord]);
@@ -1688,5 +1688,5 @@ static void build_tex_intrinsic(const struct lp_build_tgsi_action *action,
LLVMValueRef gather4_int_result_workaround = NULL;
- if (ctx->screen->info.chip_class <= VI &&
+ if (ctx->screen->info.chip_class <= GFX8 &&
opcode == TGSI_OPCODE_TG4) {
assert(inst->Texture.ReturnType != TGSI_RETURN_TYPE_UNKNOWN);
diff --git a/src/gallium/drivers/radeonsi/si_state.c b/src/gallium/drivers/radeonsi/si_state.c
index da4d30b7c9c..4823fc5c0a8 100644
--- a/src/gallium/drivers/radeonsi/si_state.c
+++ b/src/gallium/drivers/radeonsi/si_state.c
@@ -104,10 +104,10 @@ static void si_emit_cb_render_state(struct si_context *sctx)
SI_TRACKED_CB_TARGET_MASK, cb_target_mask);
- if (sctx->chip_class >= VI) {
+ if (sctx->chip_class >= GFX8) {
/* DCC MSAA workaround for blending.
* Alternatively, we can set CB_COLORi_DCC_CONTROL.OVERWRITE_-
* COMBINER_DISABLE, but that would be more complicated.
*/
- bool oc_disable = (sctx->chip_class == VI ||
+ bool oc_disable = (sctx->chip_class == GFX8 ||
sctx->chip_class == GFX9) &&
blend &&
@@ -1392,5 +1392,5 @@ static void si_emit_db_render_state(struct si_context *sctx)
bool perfect = sctx->num_perfect_occlusion_queries > 0;
- if (sctx->chip_class >= CIK) {
+ if (sctx->chip_class >= GFX7) {
unsigned log_sample_rate = sctx->framebuffer.log_samples;
@@ -1414,5 +1414,5 @@ static void si_emit_db_render_state(struct si_context *sctx)
} else {
/* Disable occlusion queries. */
- if (sctx->chip_class >= CIK) {
+ if (sctx->chip_class >= GFX7) {
db_count_control = 0;
} else {
@@ -1434,6 +1434,6 @@ static void si_emit_db_render_state(struct si_context *sctx)
db_shader_control = sctx->ps_db_shader_control;
- /* Bug workaround for smoothing (overrasterization) on SI. */
- if (sctx->chip_class == SI && sctx->smoothing_enabled) {
+ /* Bug workaround for smoothing (overrasterization) on GFX6. */
+ if (sctx->chip_class == GFX6 && sctx->smoothing_enabled) {
db_shader_control &= C_02880C_Z_ORDER;
db_shader_control |= S_02880C_Z_ORDER(V_02880C_LATE_Z);
@@ -1590,5 +1590,5 @@ static uint32_t si_translate_dbformat(enum pipe_format format)
case PIPE_FORMAT_Z24X8_UNORM:
case PIPE_FORMAT_Z24_UNORM_S8_UINT:
- return V_028040_Z_24; /* deprecated on SI */
+ return V_028040_Z_24; /* deprecated on AMD GCN */
case PIPE_FORMAT_Z32_FLOAT:
case PIPE_FORMAT_Z32_FLOAT_S8X24_UINT:
@@ -1624,7 +1624,7 @@ static uint32_t si_translate_texformat(struct pipe_screen *screen,
* Implemented as an 8_8_8_8 data format to fix texture
* gathers in stencil sampling. This affects at least
- * GL45-CTS.texture_cube_map_array.sampling on VI.
+ * GL45-CTS.texture_cube_map_array.sampling on GFX8.
*/
- if (sscreen->info.chip_class <= VI)
+ if (sscreen->info.chip_class <= GFX8)
return V_008F14_IMG_DATA_FORMAT_8_8_8_8;
@@ -2462,6 +2462,6 @@ static void si_initialize_color_surface(struct si_context *sctx,
unsigned fmask_bankh = util_logbase2(tex->surface.u.legacy.fmask.bankh);
- if (sctx->chip_class == SI) {
- /* due to a hw bug, FMASK_BANK_HEIGHT must be set on SI too */
+ if (sctx->chip_class == GFX6) {
+ /* due to a hw bug, FMASK_BANK_HEIGHT must be set on GFX6 too */
color_attrib |= S_028C74_FMASK_BANK_HEIGHT(fmask_bankh);
}
@@ -2469,5 +2469,5 @@ static void si_initialize_color_surface(struct si_context *sctx,
}
- if (sctx->chip_class >= VI) {
+ if (sctx->chip_class >= GFX8) {
unsigned max_uncompressed_block_size = V_028C78_MAX_BLOCK_SIZE_256B;
unsigned min_compressed_block_size = V_028C78_MIN_BLOCK_SIZE_32B;
@@ -2493,5 +2493,5 @@ static void si_initialize_color_surface(struct si_context *sctx,
/* This must be set for fast clear to work without FMASK. */
- if (!tex->surface.fmask_size && sctx->chip_class == SI) {
+ if (!tex->surface.fmask_size && sctx->chip_class == GFX6) {
unsigned bankh = util_logbase2(tex->surface.u.legacy.bankh);
color_attrib |= S_028C74_FMASK_BANK_HEIGHT(bankh);
@@ -2577,5 +2577,5 @@ static void si_init_depth_surface(struct si_context *sctx,
if (tex->surface.has_stencil) {
- /* Stencil buffer workaround ported from the SI-CI-VI code.
+ /* Stencil buffer workaround ported from the GFX6-GFX8 code.
* See that for explanation.
*/
@@ -2593,5 +2593,5 @@ static void si_init_depth_surface(struct si_context *sctx,
}
} else {
- /* SI-CI-VI */
+ /* GFX6-GFX8 */
struct legacy_surf_level *levelinfo = &tex->surface.u.legacy.level[level];
@@ -2608,5 +2608,5 @@ static void si_init_depth_surface(struct si_context *sctx,
surf->db_depth_info = S_02803C_ADDR5_SWIZZLE_MASK(!tex->tc_compatible_htile);
- if (sctx->chip_class >= CIK) {
+ if (sctx->chip_class >= GFX7) {
struct radeon_info *info = &sctx->screen->info;
unsigned index = tex->surface.u.legacy.tiling_index[level];
@@ -2747,5 +2747,5 @@ static void si_set_framebuffer_state(struct pipe_context *ctx,
int i;
- /* Reject zero-sized framebuffers due to a hw bug on SI that occurs
+ /* Reject zero-sized framebuffers due to a hw bug on GFX6 that occurs
* when PA_SU_HARDWARE_SCREEN_OFFSET != 0 and any_scissor.BR_X/Y <= 0.
* We could implement the full workaround here, but it's a useless case.
@@ -2936,5 +2936,5 @@ static void si_set_framebuffer_state(struct pipe_context *ctx,
/* For optimal DCC performance. */
- if (sctx->chip_class == VI)
+ if (sctx->chip_class == GFX8)
sctx->framebuffer.dcc_overwrite_combiner_watermark = 4;
else if (num_bpp64_colorbufs >= 5)
@@ -3140,5 +3140,5 @@ static void si_emit_framebuffer_state(struct si_context *sctx)
S_0287A0_EPITCH(tex->surface.u.gfx9.surf.epitch));
} else {
- /* Compute mutable surface parameters (SI-CI-VI). */
+ /* Compute mutable surface parameters (GFX6-GFX8). */
const struct legacy_surf_level *level_info =
&tex->surface.u.legacy.level[cb->base.u.tex.level];
@@ -3168,5 +3168,5 @@ static void si_emit_framebuffer_state(struct si_context *sctx)
if (tex->surface.fmask_size) {
- if (sctx->chip_class >= CIK)
+ if (sctx->chip_class >= GFX7)
cb_color_pitch |= S_028C64_FMASK_TILE_MAX(tex->surface.u.legacy.fmask.pitch_in_pixels / 8 - 1);
cb_color_attrib |= S_028C74_FMASK_TILE_MODE_INDEX(tex->surface.u.legacy.fmask.tiling_index);
@@ -3174,5 +3174,5 @@ static void si_emit_framebuffer_state(struct si_context *sctx)
} else {
/* This must be set for fast clear to work without FMASK. */
- if (sctx->chip_class >= CIK)
+ if (sctx->chip_class >= GFX7)
cb_color_pitch |= S_028C64_FMASK_TILE_MAX(pitch_tile_max);
cb_color_attrib |= S_028C74_FMASK_TILE_MODE_INDEX(tile_mode_index);
@@ -3181,5 +3181,5 @@ static void si_emit_framebuffer_state(struct si_context *sctx)
radeon_set_context_reg_seq(cs, R_028C60_CB_COLOR0_BASE + i * 0x3C,
- sctx->chip_class >= VI ? 14 : 13);
+ sctx->chip_class >= GFX8 ? 14 : 13);
radeon_emit(cs, cb_color_base); /* CB_COLOR0_BASE */
radeon_emit(cs, cb_color_pitch); /* CB_COLOR0_PITCH */
@@ -3196,5 +3196,5 @@ static void si_emit_framebuffer_state(struct si_context *sctx)
radeon_emit(cs, tex->color_clear_value[1]); /* CB_COLOR0_CLEAR_WORD1 */
- if (sctx->chip_class >= VI) /* R_028C94_CB_COLOR0_DCC_BASE */
+ if (sctx->chip_class >= GFX8) /* R_028C94_CB_COLOR0_DCC_BASE */
radeon_emit(cs, cb_dcc_base);
}
@@ -3329,5 +3329,5 @@ static void si_emit_msaa_sample_locs(struct si_context *sctx)
* if no sample lies on the pixel boundary (-8 sample offset).
*/
- bool exclusion = sctx->chip_class >= CIK &&
+ bool exclusion = sctx->chip_class >= GFX7 &&
(!rs->multisample_enable || nr_samples != 16);
radeon_opt_set_context_reg(sctx, R_02882C_PA_SU_PRIM_FILTER_CNTL,
@@ -3607,9 +3607,9 @@ si_make_buffer_descriptor(struct si_screen *screen, struct si_resource *buf,
* instruction type, STRIDE, and SWIZZLE_ENABLE.
*
- * SI-CIK:
+ * GFX6-GFX7:
* - If STRIDE == 0, it's in byte units.
* - If STRIDE != 0, it's in units of STRIDE, used with inst.IDXEN.
*
- * VI:
+ * GFX8:
* - For SMEM and STRIDE == 0, it's in byte units.
* - For SMEM and STRIDE != 0, it's in units of STRIDE.
@@ -3634,5 +3634,5 @@ si_make_buffer_descriptor(struct si_screen *screen, struct si_resource *buf,
*/
num_records = num_records ? MAX2(num_records, stride) : 0;
- else if (screen->info.chip_class == VI)
+ else if (screen->info.chip_class == GFX8)
num_records *= stride;
@@ -3721,7 +3721,7 @@ si_make_texture_descriptor(struct si_screen *screen,
* X24S8 is implemented as an 8_8_8_8 data format, to
* fix texture gathers. This affects at least
- * GL45-CTS.texture_cube_map_array.sampling on VI.
+ * GL45-CTS.texture_cube_map_array.sampling on GFX8.
*/
- if (screen->info.chip_class <= VI)
+ if (screen->info.chip_class <= GFX8)
util_format_compose_swizzles(swizzle_wwww, state_swizzle, swizzle);
else
@@ -3817,5 +3817,5 @@ si_make_texture_descriptor(struct si_screen *screen,
(res->target == PIPE_TEXTURE_CUBE ||
res->target == PIPE_TEXTURE_CUBE_ARRAY ||
- (screen->info.chip_class <= VI &&
+ (screen->info.chip_class <= GFX8 &&
res->target == PIPE_TEXTURE_3D))) {
/* For the purpose of shader images, treat cube maps and 3D
@@ -3888,5 +3888,5 @@ si_make_texture_descriptor(struct si_screen *screen,
* bits in the first dword of sampler state.
*/
- if (screen->info.chip_class <= CIK && res->nr_samples <= 1) {
+ if (screen->info.chip_class <= GFX7 && res->nr_samples <= 1) {
if (first_level == last_level)
state[7] = C_008F30_MAX_ANISO_RATIO;
@@ -4095,5 +4095,5 @@ si_create_sampler_view_custom(struct pipe_context *ctx,
depth = texture->depth0;
- if (sctx->chip_class <= VI && force_level) {
+ if (sctx->chip_class <= GFX8 && force_level) {
assert(force_level == first_level &&
force_level == last_level);
@@ -4332,5 +4332,5 @@ static void *si_create_sampler_state(struct pipe_context *ctx,
S_008F30_ANISO_BIAS(max_aniso_ratio) |
S_008F30_DISABLE_CUBE_WRAP(!state->seamless_cube_map) |
- S_008F30_COMPAT_MODE(sctx->chip_class >= VI));
+ S_008F30_COMPAT_MODE(sctx->chip_class >= GFX8));
rstate->val[1] = (S_008F34_MIN_LOD(S_FIXED(CLAMP(state->min_lod, 0, 15), 8)) |
S_008F34_MAX_LOD(S_FIXED(CLAMP(state->max_lod, 0, 15), 8)) |
@@ -4341,7 +4341,7 @@ static void *si_create_sampler_state(struct pipe_context *ctx,
S_008F38_MIP_FILTER(si_tex_mipfilter(state->min_mip_filter)) |
S_008F38_MIP_POINT_PRECLAMP(0) |
- S_008F38_DISABLE_LSB_CEIL(sctx->chip_class <= VI) |
+ S_008F38_DISABLE_LSB_CEIL(sctx->chip_class <= GFX8) |
S_008F38_FILTER_PREC_FIX(1) |
- S_008F38_ANISO_OVERRIDE(sctx->chip_class >= VI));
+ S_008F38_ANISO_OVERRIDE(sctx->chip_class >= GFX8));
rstate->val[3] = si_translate_border_color(sctx, state, &state->border_color, false);
@@ -4538,7 +4538,7 @@ static void *si_create_vertex_elements(struct pipe_context *ctx,
/* The hardware always treats the 2-bit alpha channel as
* unsigned, so a shader workaround is needed. The affected
- * chips are VI and older except Stoney (GFX8.1).
+ * chips are GFX8 and older except Stoney (GFX8.1).
*/
- always_fix = sscreen->info.chip_class <= VI &&
+ always_fix = sscreen->info.chip_class <= GFX8 &&
sscreen->info.family != CHIP_STONEY &&
channel->type == UTIL_FORMAT_TYPE_SIGNED;
@@ -4586,5 +4586,5 @@ static void *si_create_vertex_elements(struct pipe_context *ctx,
* is nicely aligned).
*/
- bool check_alignment = log_hw_load_size >= 1 && sscreen->info.chip_class == SI;
+ bool check_alignment = log_hw_load_size >= 1 && sscreen->info.chip_class == GFX6;
bool opencode = sscreen->options.vs_fetch_always_opencode;
@@ -4811,8 +4811,8 @@ static void si_memory_barrier(struct pipe_context *ctx, unsigned flags)
if (flags & PIPE_BARRIER_INDEX_BUFFER) {
- /* Indices are read through TC L2 since VI.
+ /* Indices are read through TC L2 since GFX8.
* L1 isn't used.
*/
- if (sctx->screen->info.chip_class <= CIK)
+ if (sctx->screen->info.chip_class <= GFX7)
sctx->flags |= SI_CONTEXT_WRITEBACK_GLOBAL_L2;
}
@@ -4825,10 +4825,10 @@ static void si_memory_barrier(struct pipe_context *ctx, unsigned flags)
sctx->flags |= SI_CONTEXT_FLUSH_AND_INV_CB;
- if (sctx->chip_class <= VI)
+ if (sctx->chip_class <= GFX8)
sctx->flags |= SI_CONTEXT_WRITEBACK_GLOBAL_L2;
}
/* Indirect buffers use TC L2 on GFX9, but not older hw. */
- if (sctx->screen->info.chip_class <= VI &&
+ if (sctx->screen->info.chip_class <= GFX8 &&
flags & PIPE_BARRIER_INDIRECT_BUFFER)
sctx->flags |= SI_CONTEXT_WRITEBACK_GLOBAL_L2;
@@ -4918,5 +4918,5 @@ static void si_set_grbm_gfx_index(struct si_context *sctx,
struct si_pm4_state *pm4, unsigned value)
{
- unsigned reg = sctx->chip_class >= CIK ? R_030800_GRBM_GFX_INDEX :
+ unsigned reg = sctx->chip_class >= GFX7 ? R_030800_GRBM_GFX_INDEX :
R_00802C_GRBM_GFX_INDEX;
si_pm4_set_reg(pm4, reg, value);
@@ -4955,5 +4955,5 @@ si_write_harvested_raster_configs(struct si_context *sctx,
si_set_grbm_gfx_index(sctx, pm4, ~0);
- if (sctx->chip_class >= CIK) {
+ if (sctx->chip_class >= GFX7) {
si_pm4_set_reg(pm4, R_028354_PA_SC_RASTER_CONFIG_1, raster_config_1);
}
@@ -4974,5 +4974,5 @@ static void si_set_raster_config(struct si_context *sctx, struct si_pm4_state *p
si_pm4_set_reg(pm4, R_028350_PA_SC_RASTER_CONFIG,
raster_config);
- if (sctx->chip_class >= CIK)
+ if (sctx->chip_class >= GFX7)
si_pm4_set_reg(pm4, R_028354_PA_SC_RASTER_CONFIG_1,
raster_config_1);
@@ -4989,6 +4989,6 @@ static void si_init_config(struct si_context *sctx)
struct si_pm4_state *pm4 = CALLOC_STRUCT(si_pm4_state);
- /* SI, radeon kernel disabled CLEAR_STATE. */
- assert(has_clear_state || sscreen->info.chip_class == SI ||
+ /* GFX6, radeon kernel disabled CLEAR_STATE. */
+ assert(has_clear_state || sscreen->info.chip_class == GFX6 ||
sscreen->info.drm_major != 3);
@@ -5007,5 +5007,5 @@ static void si_init_config(struct si_context *sctx)
}
- if (sctx->chip_class <= VI)
+ if (sctx->chip_class <= GFX8)
si_set_raster_config(sctx, pm4);
@@ -5015,5 +5015,5 @@ static void si_init_config(struct si_context *sctx)
/* FIXME calculate these values somehow ??? */
- if (sctx->chip_class <= VI) {
+ if (sctx->chip_class <= GFX8) {
si_pm4_set_reg(pm4, R_028A54_VGT_GS_PER_ES, SI_GS_PER_ES);
si_pm4_set_reg(pm4, R_028A58_VGT_ES_PER_GS, 0x40);
@@ -5029,5 +5029,5 @@ static void si_init_config(struct si_context *sctx)
if (!has_clear_state)
si_pm4_set_reg(pm4, R_028AB8_VGT_VTX_CNT_EN, 0x0);
- if (sctx->chip_class < CIK)
+ if (sctx->chip_class < GFX7)
si_pm4_set_reg(pm4, R_008A14_PA_CL_ENHANCE, S_008A14_NUM_CLIP_SEQ(3) |
S_008A14_CLIP_VTX_REORDER_ENA(1));
@@ -5036,5 +5036,5 @@ static void si_init_config(struct si_context *sctx)
* I don't know why. Deduced by trial and error.
*/
- if (sctx->chip_class <= CIK) {
+ if (sctx->chip_class <= GFX7) {
si_pm4_set_reg(pm4, R_028B28_VGT_STRMOUT_DRAW_OPAQUE_OFFSET, 0);
si_pm4_set_reg(pm4, R_028204_PA_SC_WINDOW_SCISSOR_TL, S_028204_WINDOW_OFFSET_DISABLE(1));
@@ -5078,5 +5078,5 @@ static void si_init_config(struct si_context *sctx)
}
- if (sctx->chip_class >= CIK) {
+ if (sctx->chip_class >= GFX7) {
if (sctx->chip_class >= GFX9) {
si_pm4_set_reg(pm4, R_00B41C_SPI_SHADER_PGM_RSRC3_HS,
@@ -5137,5 +5137,5 @@ static void si_init_config(struct si_context *sctx)
}
- if (sctx->chip_class >= VI) {
+ if (sctx->chip_class >= GFX8) {
unsigned vgt_tess_distribution;
@@ -5160,5 +5160,5 @@ static void si_init_config(struct si_context *sctx)
si_pm4_set_reg(pm4, R_028080_TA_BC_BASE_ADDR, border_color_va >> 8);
- if (sctx->chip_class >= CIK) {
+ if (sctx->chip_class >= GFX7) {
si_pm4_set_reg(pm4, R_028084_TA_BC_BASE_ADDR_HI,
S_028084_ADDRESS(border_color_va >> 40));
diff --git a/src/gallium/drivers/radeonsi/si_state_draw.c b/src/gallium/drivers/radeonsi/si_state_draw.c
index d9dfef0a381..bcce145a260 100644
--- a/src/gallium/drivers/radeonsi/si_state_draw.c
+++ b/src/gallium/drivers/radeonsi/si_state_draw.c
@@ -79,5 +79,5 @@ static void si_emit_derived_tess_state(struct si_context *sctx,
sctx->tcs_shader.cso ? sctx->tcs_shader.cso : sctx->tes_shader.cso;
unsigned tess_uses_primid = sctx->ia_multi_vgt_param_key.u.tess_uses_prim_id;
- bool has_primid_instancing_bug = sctx->chip_class == SI &&
+ bool has_primid_instancing_bug = sctx->chip_class == GFX6 &&
sctx->screen->info.max_se == 1;
unsigned tes_sh_base = sctx->shader_pointers.sh_base[PIPE_SHADER_TESS_EVAL];
@@ -153,5 +153,5 @@ static void si_emit_derived_tess_state(struct si_context *sctx,
* use LDS for the inputs and outputs.
*
- * While CIK can use 64K per threadgroup, there is a hang on Stoney
+ * While GFX7 can use 64K per threadgroup, there is a hang on Stoney
* with 2 CUs if we use more than 32K. The closed Vulkan driver also
* uses 32K at most on all GCN chips.
@@ -186,6 +186,6 @@ static void si_emit_derived_tess_state(struct si_context *sctx,
*num_patches = (temp_verts_per_tg & ~63) / max_verts_per_patch;
- if (sctx->chip_class == SI) {
- /* SI bug workaround, related to power management. Limit LS-HS
+ if (sctx->chip_class == GFX6) {
+ /* GFX6 bug workaround, related to power management. Limit LS-HS
* threadgroups to only one wave.
*/
@@ -201,5 +201,5 @@ static void si_emit_derived_tess_state(struct si_context *sctx,
* a single instance by setting SWITCH_ON_EOI, which
* should cause IA to split instances up. However, this
- * doesn't work correctly on SI when there is no other
+ * doesn't work correctly on GFX6 when there is no other
* SE to switch to.
*/
@@ -239,5 +239,5 @@ static void si_emit_derived_tess_state(struct si_context *sctx,
lds_size = output_patch0_offset + output_patch_size * *num_patches;
- if (sctx->chip_class >= CIK) {
+ if (sctx->chip_class >= GFX7) {
assert(lds_size <= 65536);
lds_size = align(lds_size, 512) / 512;
@@ -273,5 +273,5 @@ static void si_emit_derived_tess_state(struct si_context *sctx,
/* Due to a hw bug, RSRC2_LS must be written twice with another
* LS register written in between. */
- if (sctx->chip_class == CIK && sctx->family != CHIP_HAWAII)
+ if (sctx->chip_class == GFX7 && sctx->family != CHIP_HAWAII)
radeon_set_sh_reg(cs, R_00B52C_SPI_SHADER_PGM_RSRC2_LS, ls_rsrc2);
radeon_set_sh_reg_seq(cs, R_00B528_SPI_SHADER_PGM_RSRC1_LS, 2);
@@ -298,5 +298,5 @@ static void si_emit_derived_tess_state(struct si_context *sctx,
if (sctx->last_ls_hs_config != ls_hs_config) {
- if (sctx->chip_class >= CIK) {
+ if (sctx->chip_class >= GFX7) {
radeon_set_context_reg_idx(cs, R_028B58_VGT_LS_HS_CONFIG, 2,
ls_hs_config);
@@ -350,8 +350,8 @@ si_get_init_multi_vgt_param(struct si_screen *sscreen,
partial_vs_wave = true;
- /* Needed for 028B6C_DISTRIBUTION_MODE != 0. (implies >= VI) */
+ /* Needed for 028B6C_DISTRIBUTION_MODE != 0. (implies >= GFX8) */
if (sscreen->has_distributed_tess) {
if (key->u.uses_gs) {
- if (sscreen->info.chip_class == VI)
+ if (sscreen->info.chip_class == GFX8)
partial_es_wave = true;
} else {
@@ -368,5 +368,5 @@ si_get_init_multi_vgt_param(struct si_screen *sscreen,
}
- if (sscreen->info.chip_class >= CIK) {
+ if (sscreen->info.chip_class >= GFX7) {
/* WD_SWITCH_ON_EOP has no effect on GPUs with less than
* 4 shader engines. Set 1 to pass the assertion below.
@@ -401,10 +401,10 @@ si_get_init_multi_vgt_param(struct si_screen *sscreen,
* This is needed for good VS wave utilization.
*/
- if (sscreen->info.chip_class <= VI &&
+ if (sscreen->info.chip_class <= GFX8 &&
sscreen->info.max_se == 4 &&
key->u.multi_instances_smaller_than_primgroup)
wd_switch_on_eop = true;
- /* Required on CIK and later. */
+ /* Required on GFX7 and later. */
if (sscreen->info.max_se == 4 && !wd_switch_on_eop)
ia_switch_on_eoi = true;
@@ -422,8 +422,8 @@ si_get_init_multi_vgt_param(struct si_screen *sscreen,
partial_vs_wave = true;
- /* Required by Hawaii and, for some special cases, by VI. */
+ /* Required by Hawaii and, for some special cases, by GFX8. */
if (ia_switch_on_eoi &&
(sscreen->info.family == CHIP_HAWAII ||
- (sscreen->info.chip_class == VI &&
+ (sscreen->info.chip_class == GFX8 &&
(key->u.uses_gs || max_primgroup_in_wave != 2))))
partial_vs_wave = true;
@@ -445,5 +445,5 @@ si_get_init_multi_vgt_param(struct si_screen *sscreen,
/* If SWITCH_ON_EOI is set, PARTIAL_ES_WAVE must be set too. */
- if (sscreen->info.chip_class <= VI && ia_switch_on_eoi)
+ if (sscreen->info.chip_class <= GFX8 && ia_switch_on_eoi)
partial_es_wave = true;
@@ -452,7 +452,7 @@ si_get_init_multi_vgt_param(struct si_screen *sscreen,
S_028AA8_PARTIAL_VS_WAVE_ON(partial_vs_wave) |
S_028AA8_PARTIAL_ES_WAVE_ON(partial_es_wave) |
- S_028AA8_WD_SWITCH_ON_EOP(sscreen->info.chip_class >= CIK ? wd_switch_on_eop : 0) |
+ S_028AA8_WD_SWITCH_ON_EOP(sscreen->info.chip_class >= GFX7 ? wd_switch_on_eop : 0) |
/* The following field was moved to VGT_SHADER_STAGES_EN in GFX9. */
- S_028AA8_MAX_PRIMGRP_IN_WAVE(sscreen->info.chip_class == VI ?
+ S_028AA8_MAX_PRIMGRP_IN_WAVE(sscreen->info.chip_class == GFX8 ?
max_primgroup_in_wave : 0) |
S_030960_EN_INST_OPT_BASIC(sscreen->info.chip_class >= GFX9) |
@@ -520,5 +520,5 @@ static unsigned si_get_ia_multi_vgt_param(struct si_context *sctx,
if (sctx->gs_shader.cso) {
/* GS requirement. */
- if (sctx->chip_class <= VI &&
+ if (sctx->chip_class <= GFX8 &&
SI_GS_PER_ES / primgroup_size >= sctx->screen->gs_table_depth - 3)
ia_multi_vgt_param |= S_028AA8_PARTIAL_ES_WAVE_ON(1);
@@ -626,5 +626,5 @@ static void si_emit_draw_registers(struct si_context *sctx,
R_030960_IA_MULTI_VGT_PARAM, 4,
ia_multi_vgt_param);
- else if (sctx->chip_class >= CIK)
+ else if (sctx->chip_class >= GFX7)
radeon_set_context_reg_idx(cs, R_028AA8_IA_MULTI_VGT_PARAM, 1, ia_multi_vgt_param);
else
@@ -634,5 +634,5 @@ static void si_emit_draw_registers(struct si_context *sctx,
}
if (prim != sctx->last_prim) {
- if (sctx->chip_class >= CIK)
+ if (sctx->chip_class >= GFX7)
radeon_set_uconfig_reg_idx(cs, sctx->screen,
R_030908_VGT_PRIMITIVE_TYPE, 1, prim);
@@ -701,10 +701,10 @@ static void si_emit_draw_packets(struct si_context *sctx,
case 2:
index_type = V_028A7C_VGT_INDEX_16 |
- (SI_BIG_ENDIAN && sctx->chip_class <= CIK ?
+ (SI_BIG_ENDIAN && sctx->chip_class <= GFX7 ?
V_028A7C_VGT_DMA_SWAP_16_BIT : 0);
break;
case 4:
index_type = V_028A7C_VGT_INDEX_32 |
- (SI_BIG_ENDIAN && sctx->chip_class <= CIK ?
+ (SI_BIG_ENDIAN && sctx->chip_class <= GFX7 ?
V_028A7C_VGT_DMA_SWAP_32_BIT : 0);
break;
@@ -734,8 +734,8 @@ static void si_emit_draw_packets(struct si_context *sctx,
RADEON_USAGE_READ, RADEON_PRIO_INDEX_BUFFER);
} else {
- /* On CI and later, non-indexed draws overwrite VGT_INDEX_TYPE,
+ /* On GFX7 and later, non-indexed draws overwrite VGT_INDEX_TYPE,
* so the state must be re-emitted before the next indexed draw.
*/
- if (sctx->chip_class >= CIK)
+ if (sctx->chip_class >= GFX7)
sctx->last_index_size = -1;
}
@@ -919,5 +919,5 @@ void si_emit_cache_flush(struct si_context *sctx)
sctx->num_db_cache_flushes++;
- /* SI has a bug that it always flushes ICACHE and KCACHE if either
+ /* GFX6 has a bug that it always flushes ICACHE and KCACHE if either
* bit is set. An alternative way is to write SQC_CACHES, but that
* doesn't seem to work reliably. Since the bug doesn't affect
@@ -932,5 +932,5 @@ void si_emit_cache_flush(struct si_context *sctx)
cp_coher_cntl |= S_0085F0_SH_KCACHE_ACTION_ENA(1);
- if (sctx->chip_class <= VI) {
+ if (sctx->chip_class <= GFX8) {
if (flags & SI_CONTEXT_FLUSH_AND_INV_CB) {
cp_coher_cntl |= S_0085F0_CB_ACTION_ENA(1) |
@@ -945,5 +945,5 @@ void si_emit_cache_flush(struct si_context *sctx)
/* Necessary for DCC */
- if (sctx->chip_class == VI)
+ if (sctx->chip_class == GFX8)
si_cp_release_mem(sctx,
V_028A90_FLUSH_AND_INV_CB_DATA_TS,
@@ -1086,5 +1086,5 @@ void si_emit_cache_flush(struct si_context *sctx)
}
- /* SI-CI-VI only:
+ /* GFX6-GFX8 only:
* When one of the CP_COHER_CNTL.DEST_BASE flags is set, SURFACE_SYNC
* waits for idle, so it should be last. SURFACE_SYNC is done in PFP.
@@ -1093,16 +1093,16 @@ void si_emit_cache_flush(struct si_context *sctx)
* at this point.
*
- * SI-CIK don't support L2 write-back.
+ * GFX6-GFX7 don't support L2 write-back.
*/
if (flags & SI_CONTEXT_INV_GLOBAL_L2 ||
- (sctx->chip_class <= CIK &&
+ (sctx->chip_class <= GFX7 &&
(flags & SI_CONTEXT_WRITEBACK_GLOBAL_L2))) {
- /* Invalidate L1 & L2. (L1 is always invalidated on SI)
- * WB must be set on VI+ when TC_ACTION is set.
+ /* Invalidate L1 & L2. (L1 is always invalidated on GFX6)
+ * WB must be set on GFX8+ when TC_ACTION is set.
*/
si_emit_surface_sync(sctx, cp_coher_cntl |
S_0085F0_TC_ACTION_ENA(1) |
S_0085F0_TCL1_ACTION_ENA(1) |
- S_0301F0_TC_WB_ACTION_ENA(sctx->chip_class >= VI));
+ S_0301F0_TC_WB_ACTION_ENA(sctx->chip_class >= GFX8));
cp_coher_cntl = 0;
sctx->num_L2_invalidates++;
@@ -1261,5 +1261,5 @@ static void si_draw_vbo(struct pipe_context *ctx, const struct pipe_draw_info *i
if (likely(!info->indirect)) {
- /* SI-CI treat instance_count==0 as instance_count==1. There is
+ /* GFX6-GFX7 treat instance_count==0 as instance_count==1. There is
* no workaround for indirect draws, but we can at least skip
* direct draws.
@@ -1370,6 +1370,6 @@ static void si_draw_vbo(struct pipe_context *ctx, const struct pipe_draw_info *i
if (index_size) {
/* Translate or upload, if needed. */
- /* 8-bit indices are supported on VI. */
- if (sctx->chip_class <= CIK && index_size == 1) {
+ /* 8-bit indices are supported on GFX8. */
+ if (sctx->chip_class <= GFX7 && index_size == 1) {
unsigned start, count, start_offset, size, offset;
void *ptr;
@@ -1411,7 +1411,7 @@ static void si_draw_vbo(struct pipe_context *ctx, const struct pipe_draw_info *i
/* info->start will be added by the drawing code */
index_offset -= start_offset;
- } else if (sctx->chip_class <= CIK &&
+ } else if (sctx->chip_class <= GFX7 &&
si_resource(indexbuf)->TC_L2_dirty) {
- /* VI reads index buffers through TC L2, so it doesn't
+ /* GFX8 reads index buffers through TC L2, so it doesn't
* need this. */
sctx->flags |= SI_CONTEXT_WRITEBACK_GLOBAL_L2;
@@ -1427,5 +1427,5 @@ static void si_draw_vbo(struct pipe_context *ctx, const struct pipe_draw_info *i
/* Indirect buffers use TC L2 on GFX9, but not older hw. */
- if (sctx->chip_class <= VI) {
+ if (sctx->chip_class <= GFX8) {
if (si_resource(indirect->buffer)->TC_L2_dirty) {
sctx->flags |= SI_CONTEXT_WRITEBACK_GLOBAL_L2;
@@ -1506,5 +1506,5 @@ static void si_draw_vbo(struct pipe_context *ctx, const struct pipe_draw_info *i
* in parallel, but starting the draw first is more important.
*/
- if (sctx->chip_class >= CIK && sctx->prefetch_L2_mask)
+ if (sctx->chip_class >= GFX7 && sctx->prefetch_L2_mask)
cik_emit_prefetch_L2(sctx, false);
} else {
@@ -1516,5 +1516,5 @@ static void si_draw_vbo(struct pipe_context *ctx, const struct pipe_draw_info *i
/* Only prefetch the API VS and VBO descriptors. */
- if (sctx->chip_class >= CIK && sctx->prefetch_L2_mask)
+ if (sctx->chip_class >= GFX7 && sctx->prefetch_L2_mask)
cik_emit_prefetch_L2(sctx, true);
@@ -1535,5 +1535,5 @@ static void si_draw_vbo(struct pipe_context *ctx, const struct pipe_draw_info *i
/* Prefetch the remaining shaders after the draw has been
* started. */
- if (sctx->chip_class >= CIK && sctx->prefetch_L2_mask)
+ if (sctx->chip_class >= GFX7 && sctx->prefetch_L2_mask)
cik_emit_prefetch_L2(sctx, false);
}
diff --git a/src/gallium/drivers/radeonsi/si_state_shaders.c b/src/gallium/drivers/radeonsi/si_state_shaders.c
index 51a3af92d0c..10677f175de 100644
--- a/src/gallium/drivers/radeonsi/si_state_shaders.c
+++ b/src/gallium/drivers/radeonsi/si_state_shaders.c
@@ -474,5 +474,5 @@ static void si_shader_ls(struct si_screen *sscreen, struct si_shader *shader)
uint64_t va;
- assert(sscreen->info.chip_class <= VI);
+ assert(sscreen->info.chip_class <= GFX8);
pm4 = si_get_shader_pm4_state(shader);
@@ -548,5 +548,5 @@ static void si_shader_hs(struct si_screen *sscreen, struct si_shader *shader)
S_00B428_LS_VGPR_COMP_CNT(ls_vgpr_comp_cnt));
- if (sscreen->info.chip_class <= VI) {
+ if (sscreen->info.chip_class <= GFX8) {
si_pm4_set_reg(pm4, R_00B42C_SPI_SHADER_PGM_RSRC2_HS,
shader->config.rsrc2);
@@ -588,5 +588,5 @@ static void si_shader_es(struct si_screen *sscreen, struct si_shader *shader)
unsigned oc_lds_en;
- assert(sscreen->info.chip_class <= VI);
+ assert(sscreen->info.chip_class <= GFX8);
pm4 = si_get_shader_pm4_state(shader);
@@ -974,5 +974,5 @@ static void si_emit_shader_vs(struct si_context *sctx)
shader->ctx_reg.vs.vgt_primitiveid_en);
- if (sctx->chip_class <= VI) {
+ if (sctx->chip_class <= GFX8) {
radeon_opt_set_context_reg(sctx, R_028AB4_VGT_REUSE_OFF,
SI_TRACKED_VGT_REUSE_OFF,
@@ -1053,5 +1053,5 @@ static void si_shader_vs(struct si_screen *sscreen, struct si_shader *shader,
}
- if (sscreen->info.chip_class <= VI) {
+ if (sscreen->info.chip_class <= GFX8) {
/* Reuse needs to be set off if we write oViewport. */
shader->ctx_reg.vs.vgt_reuse_off =
@@ -1603,9 +1603,9 @@ static inline void si_shader_selector_key(struct pipe_context *ctx,
key->part.ps.epilog.spi_shader_col_format |= V_028710_SPI_SHADER_32_AR;
- /* On SI and CIK except Hawaii, the CB doesn't clamp outputs
+ /* On GFX6 and GFX7 except Hawaii, the CB doesn't clamp outputs
* to the range supported by the type if a channel has less
* than 16 bits and the export format is 16_ABGR.
*/
- if (sctx->chip_class <= CIK && sctx->family != CHIP_HAWAII) {
+ if (sctx->chip_class <= GFX7 && sctx->family != CHIP_HAWAII) {
key->part.ps.epilog.color_is_int8 = sctx->framebuffer.color_is_int8;
key->part.ps.epilog.color_is_int10 = sctx->framebuffer.color_is_int10;
@@ -2707,8 +2707,8 @@ static void si_delete_shader(struct si_context *sctx, struct si_shader *shader)
case PIPE_SHADER_VERTEX:
if (shader->key.as_ls) {
- assert(sctx->chip_class <= VI);
+ assert(sctx->chip_class <= GFX8);
si_pm4_delete_state(sctx, ls, shader->pm4);
} else if (shader->key.as_es) {
- assert(sctx->chip_class <= VI);
+ assert(sctx->chip_class <= GFX8);
si_pm4_delete_state(sctx, es, shader->pm4);
} else {
@@ -2721,5 +2721,5 @@ static void si_delete_shader(struct si_context *sctx, struct si_shader *shader)
case PIPE_SHADER_TESS_EVAL:
if (shader->key.as_es) {
- assert(sctx->chip_class <= VI);
+ assert(sctx->chip_class <= GFX8);
si_pm4_delete_state(sctx, es, shader->pm4);
} else {
@@ -2938,8 +2938,8 @@ static bool si_update_gs_ring_buffers(struct si_context *sctx)
unsigned wave_size = 64;
unsigned max_gs_waves = 32 * num_se; /* max 32 per SE on GCN */
- /* On SI-CI, the value comes from VGT_GS_VERTEX_REUSE = 16.
- * On VI+, the value comes from VGT_VERTEX_REUSE_BLOCK_CNTL = 30 (+2).
+ /* On GFX6-GFX7, the value comes from VGT_GS_VERTEX_REUSE = 16.
+ * On GFX8+, the value comes from VGT_VERTEX_REUSE_BLOCK_CNTL = 30 (+2).
*/
- unsigned gs_vertex_reuse = (sctx->chip_class >= VI ? 32 : 16) * num_se;
+ unsigned gs_vertex_reuse = (sctx->chip_class >= GFX8 ? 32 : 16) * num_se;
unsigned alignment = 256 * num_se;
/* The maximum size is 63.999 MB per SE. */
@@ -2968,5 +2968,5 @@ static bool si_update_gs_ring_buffers(struct si_context *sctx)
* GFX9 doesn't have the ESGS ring.
*/
- bool update_esgs = sctx->chip_class <= VI &&
+ bool update_esgs = sctx->chip_class <= GFX8 &&
esgs_ring_size &&
(!sctx->esgs_ring ||
@@ -3006,7 +3006,7 @@ static bool si_update_gs_ring_buffers(struct si_context *sctx)
return false;
- if (sctx->chip_class >= CIK) {
+ if (sctx->chip_class >= GFX7) {
if (sctx->esgs_ring) {
- assert(sctx->chip_class <= VI);
+ assert(sctx->chip_class <= GFX8);
si_pm4_set_reg(pm4, R_030900_VGT_ESGS_RING_SIZE,
sctx->esgs_ring->width0 / 256);
@@ -3040,5 +3040,5 @@ static bool si_update_gs_ring_buffers(struct si_context *sctx)
/* Set ring bindings. */
if (sctx->esgs_ring) {
- assert(sctx->chip_class <= VI);
+ assert(sctx->chip_class <= GFX8);
si_set_ring_buffer(sctx, SI_ES_RING_ESGS,
sctx->esgs_ring, 0, sctx->esgs_ring->width0,
@@ -3289,5 +3289,5 @@ static void si_init_tess_factor_ring(struct si_context *sctx)
/* Append these registers to the init config state. */
- if (sctx->chip_class >= CIK) {
+ if (sctx->chip_class >= GFX7) {
si_pm4_set_reg(sctx->init_config, R_030938_VGT_TF_RING_SIZE,
S_030938_SIZE(sctx->screen->tess_factor_ring_size / 4));
@@ -3377,5 +3377,5 @@ bool si_update_shaders(struct si_context *sctx)
/* VS as LS */
- if (sctx->chip_class <= VI) {
+ if (sctx->chip_class <= GFX8) {
r = si_shader_select(ctx, &sctx->vs_shader,
&compiler_state);
@@ -3409,5 +3409,5 @@ bool si_update_shaders(struct si_context *sctx)
if (sctx->gs_shader.cso) {
/* TES as ES */
- if (sctx->chip_class <= VI) {
+ if (sctx->chip_class <= GFX8) {
r = si_shader_select(ctx, &sctx->tes_shader,
&compiler_state);
@@ -3425,5 +3425,5 @@ bool si_update_shaders(struct si_context *sctx)
}
} else if (sctx->gs_shader.cso) {
- if (sctx->chip_class <= VI) {
+ if (sctx->chip_class <= GFX8) {
/* VS as ES */
r = si_shader_select(ctx, &sctx->vs_shader,
@@ -3458,5 +3458,5 @@ bool si_update_shaders(struct si_context *sctx)
} else {
si_pm4_bind_state(sctx, gs, NULL);
- if (sctx->chip_class <= VI)
+ if (sctx->chip_class <= GFX8)
si_pm4_bind_state(sctx, es, NULL);
}
@@ -3505,5 +3505,5 @@ bool si_update_shaders(struct si_context *sctx)
si_mark_atom_dirty(sctx, &sctx->atoms.s.msaa_config);
- if (sctx->chip_class == SI)
+ if (sctx->chip_class == GFX6)
si_mark_atom_dirty(sctx, &sctx->atoms.s.db_render_state);
@@ -3523,5 +3523,5 @@ bool si_update_shaders(struct si_context *sctx)
}
- if (sctx->chip_class >= CIK) {
+ if (sctx->chip_class >= GFX7) {
if (si_pm4_state_enabled_and_changed(sctx, ls))
sctx->prefetch_L2_mask |= SI_PREFETCH_LS;
diff --git a/src/gallium/drivers/radeonsi/si_state_streamout.c b/src/gallium/drivers/radeonsi/si_state_streamout.c
index 2a0a4bef9a2..e7058f19a8a 100644
--- a/src/gallium/drivers/radeonsi/si_state_streamout.c
+++ b/src/gallium/drivers/radeonsi/si_state_streamout.c
@@ -104,5 +104,5 @@ static void si_set_streamout_targets(struct pipe_context *ctx,
*
* The only cases which requires flushing it is VGT DMA index
- * fetching (on <= CIK) and indirect draw data, which are rare
+ * fetching (on <= GFX7) and indirect draw data, which are rare
* cases. Thus, flag the TC L2 dirtiness in the resource and
* handle it at draw call time.
@@ -196,5 +196,5 @@ static void si_flush_vgt_streamout(struct si_context *sctx)
/* The register is at different places on different ASICs. */
- if (sctx->chip_class >= CIK) {
+ if (sctx->chip_class >= GFX7) {
reg_strmout_cntl = R_0300FC_CP_STRMOUT_CNTL;
radeon_set_uconfig_reg(cs, reg_strmout_cntl, 0);
@@ -231,5 +231,5 @@ static void si_emit_streamout_begin(struct si_context *sctx)
t[i]->stride_in_dw = stride_in_dw[i];
- /* SI binds streamout buffers as shader resources.
+ /* AMD GCN binds streamout buffers as shader resources.
* VGT only counts primitives and tells the shader
* through SGPRs what to do. */
diff --git a/src/gallium/drivers/radeonsi/si_state_viewport.c b/src/gallium/drivers/radeonsi/si_state_viewport.c
index 792d1c4efd1..a144d7b661c 100644
--- a/src/gallium/drivers/radeonsi/si_state_viewport.c
+++ b/src/gallium/drivers/radeonsi/si_state_viewport.c
@@ -127,8 +127,8 @@ static void si_emit_one_scissor(struct si_context *ctx,
si_clip_scissor(&final, scissor);
- /* Workaround for a hw bug on SI that occurs when PA_SU_HARDWARE_-
+ /* Workaround for a hw bug on GFX6 that occurs when PA_SU_HARDWARE_-
* SCREEN_OFFSET != 0 and any_scissor.BR_X/Y <= 0.
*/
- if (ctx->chip_class == SI && (final.maxx == 0 || final.maxy == 0)) {
+ if (ctx->chip_class == GFX6 && (final.maxx == 0 || final.maxy == 0)) {
radeon_emit(cs, S_028250_TL_X(1) |
S_028250_TL_Y(1) |
@@ -181,7 +181,7 @@ static void si_emit_guardband(struct si_context *ctx)
int hw_screen_offset_y = (vp_as_scissor.maxy + vp_as_scissor.miny) / 2;
- /* SI-CI need to align the offset to an ubertile consisting of all SEs. */
+ /* GFX6-GFX7 need to align the offset to an ubertile consisting of all SEs. */
const unsigned hw_screen_offset_alignment =
- ctx->chip_class >= VI ? 16 : MAX2(ctx->screen->se_tile_repeat, 16);
+ ctx->chip_class >= GFX8 ? 16 : MAX2(ctx->screen->se_tile_repeat, 16);
/* Indexed by quantization modes */
diff --git a/src/gallium/drivers/radeonsi/si_test_dma_perf.c b/src/gallium/drivers/radeonsi/si_test_dma_perf.c
index 124f5bb5c12..263187d683f 100644
--- a/src/gallium/drivers/radeonsi/si_test_dma_perf.c
+++ b/src/gallium/drivers/radeonsi/si_test_dma_perf.c
@@ -113,9 +113,9 @@ void si_test_dma_perf(struct si_screen *sscreen)
test_cs ? cs_dwords_per_thread_list[cs_method % NUM_SHADERS] : 0;
- if (sctx->chip_class == SI) {
- /* SI doesn't support CP DMA operations through L2. */
+ if (sctx->chip_class == GFX6) {
+ /* GFX6 doesn't support CP DMA operations through L2. */
if (test_cp && cache_policy != L2_BYPASS)
continue;
- /* WAVES_PER_SH is in multiples of 16 on SI. */
+ /* WAVES_PER_SH is in multiples of 16 on GFX6. */
if (test_cs && cs_waves_per_sh % 16 != 0)
continue;
@@ -152,5 +152,5 @@ void si_test_dma_perf(struct si_screen *sscreen)
if (test_sdma) {
- if (sctx->chip_class == SI)
+ if (sctx->chip_class == GFX6)
query_type = SI_QUERY_TIME_ELAPSED_SDMA_SI;
else
@@ -347,8 +347,8 @@ void si_test_dma_perf(struct si_screen *sscreen)
continue;
- /* Ban CP DMA clears via MC on <= VI. They are super slow
+ /* Ban CP DMA clears via MC on <= GFX8. They are super slow
* on GTT, which we can get due to BO evictions.
*/
- if (sctx->chip_class <= VI && placement == 1 &&
+ if (sctx->chip_class <= GFX8 && placement == 1 &&
r->is_cp && r->cache_policy == L2_BYPASS)
continue;
diff --git a/src/gallium/drivers/radeonsi/si_texture.c b/src/gallium/drivers/radeonsi/si_texture.c
index 59d50376438..74c9cf9d7bf 100644
--- a/src/gallium/drivers/radeonsi/si_texture.c
+++ b/src/gallium/drivers/radeonsi/si_texture.c
@@ -255,8 +255,8 @@ static int si_init_surface(struct si_screen *sscreen,
/* TC-compatible HTILE only supports Z32_FLOAT.
* GFX9 also supports Z16_UNORM.
- * On VI, promote Z16 to Z32. DB->CB copies will convert
+ * On GFX8, promote Z16 to Z32. DB->CB copies will convert
* the format for transfers.
*/
- if (sscreen->info.chip_class == VI)
+ if (sscreen->info.chip_class == GFX8)
bpe = 4;
@@ -268,5 +268,5 @@ static int si_init_surface(struct si_screen *sscreen,
}
- if (sscreen->info.chip_class >= VI &&
+ if (sscreen->info.chip_class >= GFX8 &&
(ptex->flags & SI_RESOURCE_FLAG_DISABLE_DCC ||
ptex->format == PIPE_FORMAT_R9G9B9E5_FLOAT ||
@@ -279,6 +279,6 @@ static int si_init_surface(struct si_screen *sscreen,
flags |= RADEON_SURF_DISABLE_DCC;
- /* VI: DCC clear for 4x and 8x MSAA array textures unimplemented. */
- if (sscreen->info.chip_class == VI &&
+ /* GFX8: DCC clear for 4x and 8x MSAA array textures unimplemented. */
+ if (sscreen->info.chip_class == GFX8 &&
ptex->nr_storage_samples >= 4 &&
ptex->array_size > 1)
@@ -701,5 +701,5 @@ static void si_set_tex_bo_metadata(struct si_screen *sscreen,
/* Dwords [10:..] contain the mipmap level offsets. */
- if (sscreen->info.chip_class <= VI) {
+ if (sscreen->info.chip_class <= GFX8) {
for (unsigned i = 0; i <= res->last_level; i++)
md.metadata[10+i] = tex->surface.u.legacy.level[i].offset >> 8;
@@ -717,5 +717,5 @@ static void si_get_opaque_metadata(struct si_screen *sscreen,
uint32_t *desc = &md->metadata[2];
- if (sscreen->info.chip_class < VI)
+ if (sscreen->info.chip_class < GFX8)
return;
@@ -758,5 +758,5 @@ static bool si_has_displayable_dcc(struct si_texture *tex)
struct si_screen *sscreen = (struct si_screen*)tex->buffer.b.b.screen;
- if (sscreen->info.chip_class <= VI)
+ if (sscreen->info.chip_class <= GFX8)
return false;
@@ -850,5 +850,5 @@ static boolean si_texture_get_handle(struct pipe_screen* screen,
}
- /* Since shader image stores don't support DCC on VI,
+ /* Since shader image stores don't support DCC on GFX8,
* disable it for external clients that want write
* access.
@@ -975,5 +975,5 @@ static void si_texture_get_htile_size(struct si_screen *sscreen,
unsigned num_pipes = sscreen->info.num_tile_pipes;
- assert(sscreen->info.chip_class <= VI);
+ assert(sscreen->info.chip_class <= GFX8);
tex->surface.htile_size = 0;
@@ -990,5 +990,5 @@ static void si_texture_get_htile_size(struct si_screen *sscreen,
* on Carrizo too, though it was very rare there.
*/
- if (sscreen->info.chip_class >= CIK && num_pipes < 4)
+ if (sscreen->info.chip_class >= GFX7 && num_pipes < 4)
num_pipes = 4;
@@ -1037,5 +1037,5 @@ static void si_texture_allocate_htile(struct si_screen *sscreen,
struct si_texture *tex)
{
- if (sscreen->info.chip_class <= VI && !tex->tc_compatible_htile)
+ if (sscreen->info.chip_class <= GFX8 && !tex->tc_compatible_htile)
si_texture_get_htile_size(sscreen, tex);
@@ -1230,5 +1230,5 @@ si_texture_create_object(struct pipe_screen *screen,
/* TC-compatible HTILE:
- * - VI only supports Z32_FLOAT.
+ * - GFX8 only supports Z32_FLOAT.
* - GFX9 only supports Z32_FLOAT and Z16_UNORM. */
if (tex->tc_compatible_htile) {
@@ -1507,8 +1507,8 @@ si_choose_tiling(struct si_screen *sscreen,
return RADEON_SURF_MODE_LINEAR_ALIGNED;
- /* Avoid Z/S decompress blits by forcing TC-compatible HTILE on VI,
+ /* Avoid Z/S decompress blits by forcing TC-compatible HTILE on GFX8,
* which requires 2D tiling.
*/
- if (sscreen->info.chip_class == VI && tc_compatible_htile)
+ if (sscreen->info.chip_class == GFX8 && tc_compatible_htile)
return RADEON_SURF_MODE_2D;
@@ -1526,5 +1526,5 @@ si_choose_tiling(struct si_screen *sscreen,
return RADEON_SURF_MODE_LINEAR_ALIGNED;
- /* Cursors are linear on SI.
+ /* Cursors are linear on AMD GCN.
* (XXX double-check, maybe also use RADEON_SURF_SCANOUT) */
if (templ->bind & PIPE_BIND_CURSOR)
@@ -1583,5 +1583,5 @@ struct pipe_resource *si_texture_create(struct pipe_screen *screen,
bool is_flushed_depth = templ->flags & SI_RESOURCE_FLAG_FLUSHED_DEPTH;
bool tc_compatible_htile =
- sscreen->info.chip_class >= VI &&
+ sscreen->info.chip_class >= GFX8 &&
/* There are issues with TC-compatible HTILE on Tonga (and
* Iceland is the same design), and documented bug workarounds
@@ -2451,5 +2451,5 @@ void vi_separate_dcc_try_enable(struct si_context *sctx,
return;
- assert(sctx->chip_class >= VI);
+ assert(sctx->chip_class >= GFX8);
if (tex->dcc_offset)
diff --git a/src/gallium/winsys/amdgpu/drm/amdgpu_cs.c b/src/gallium/winsys/amdgpu/drm/amdgpu_cs.c
index ad663e7d5aa..0f5890baecd 100644
--- a/src/gallium/winsys/amdgpu/drm/amdgpu_cs.c
+++ b/src/gallium/winsys/amdgpu/drm/amdgpu_cs.c
@@ -384,5 +384,5 @@ static bool amdgpu_cs_has_user_fence(struct amdgpu_cs_context *cs)
static bool amdgpu_cs_has_chaining(struct amdgpu_cs *cs)
{
- return cs->ctx->ws->info.chip_class >= CIK &&
+ return cs->ctx->ws->info.chip_class >= GFX7 &&
(cs->ring_type == RING_GFX || cs->ring_type == RING_COMPUTE);
}
@@ -1523,5 +1523,5 @@ static int amdgpu_cs_flush(struct radeon_cmdbuf *rcs,
case RING_DMA:
/* pad DMA ring to 8 DWs */
- if (ws->info.chip_class <= SI) {
+ if (ws->info.chip_class <= GFX6) {
while (rcs->current.cdw & 7)
radeon_emit(rcs, 0xf0000000); /* NOP packet */
diff --git a/src/gallium/winsys/amdgpu/drm/amdgpu_winsys.c b/src/gallium/winsys/amdgpu/drm/amdgpu_winsys.c
index cf02311864e..1430c94d3e4 100644
--- a/src/gallium/winsys/amdgpu/drm/amdgpu_winsys.c
+++ b/src/gallium/winsys/amdgpu/drm/amdgpu_winsys.c
@@ -69,9 +69,9 @@ static void handle_env_var_force_family(struct amdgpu_winsys *ws)
ws->info.chip_class = GFX9;
else if (i >= CHIP_TONGA)
- ws->info.chip_class = VI;
+ ws->info.chip_class = GFX8;
else if (i >= CHIP_BONAIRE)
- ws->info.chip_class = CIK;
+ ws->info.chip_class = GFX7;
else
- ws->info.chip_class = SI;
+ ws->info.chip_class = GFX6;
/* Don't submit any IBs. */
diff --git a/src/gallium/winsys/radeon/drm/radeon_drm_cs.c b/src/gallium/winsys/radeon/drm/radeon_drm_cs.c
index 02c10f7ea7d..4032944d0c8 100644
--- a/src/gallium/winsys/radeon/drm/radeon_drm_cs.c
+++ b/src/gallium/winsys/radeon/drm/radeon_drm_cs.c
@@ -571,5 +571,5 @@ static int radeon_drm_cs_flush(struct radeon_cmdbuf *rcs,
case RING_DMA:
/* pad DMA ring to 8 DWs */
- if (cs->ws->info.chip_class <= SI) {
+ if (cs->ws->info.chip_class <= GFX6) {
while (rcs->current.cdw & 7)
radeon_emit(&cs->base, 0xf0000000); /* NOP packet */
diff --git a/src/gallium/winsys/radeon/drm/radeon_drm_surface.c b/src/gallium/winsys/radeon/drm/radeon_drm_surface.c
index 20cfc86ebe0..d33c4c7132d 100644
--- a/src/gallium/winsys/radeon/drm/radeon_drm_surface.c
+++ b/src/gallium/winsys/radeon/drm/radeon_drm_surface.c
@@ -51,5 +51,5 @@ static void set_micro_tile_mode(struct radeon_surf *surf,
uint32_t tile_mode;
- if (info->chip_class < SI) {
+ if (info->chip_class < GFX6) {
surf->micro_tile_mode = 0;
return;
@@ -58,5 +58,5 @@ static void set_micro_tile_mode(struct radeon_surf *surf,
tile_mode = info->si_tile_mode_array[surf->u.legacy.tiling_index[0]];
- if (info->chip_class >= CIK)
+ if (info->chip_class >= GFX7)
surf->micro_tile_mode = G_009910_MICRO_TILE_MODE_NEW(tile_mode);
else
@@ -232,5 +232,5 @@ static void si_compute_cmask(const struct radeon_info *info,
return;
- assert(info->chip_class <= VI);
+ assert(info->chip_class <= GFX8);
switch (num_pipes) {
diff --git a/src/gallium/winsys/radeon/drm/radeon_drm_winsys.c b/src/gallium/winsys/radeon/drm/radeon_drm_winsys.c
index de20edbe662..99ecbea283b 100644
--- a/src/gallium/winsys/radeon/drm/radeon_drm_winsys.c
+++ b/src/gallium/winsys/radeon/drm/radeon_drm_winsys.c
@@ -270,5 +270,5 @@ static bool do_winsys_init(struct radeon_drm_winsys *ws)
case CHIP_OLAND:
case CHIP_HAINAN:
- ws->info.chip_class = SI;
+ ws->info.chip_class = GFX6;
break;
case CHIP_BONAIRE:
@@ -277,5 +277,5 @@ static bool do_winsys_init(struct radeon_drm_winsys *ws)
case CHIP_HAWAII:
case CHIP_MULLINS:
- ws->info.chip_class = CIK;
+ ws->info.chip_class = GFX7;
break;
}
@@ -543,16 +543,16 @@ static bool do_winsys_init(struct radeon_drm_winsys *ws)
}
- if (ws->info.chip_class == CIK) {
+ if (ws->info.chip_class == GFX7) {
if (!radeon_get_drm_value(ws->fd, RADEON_INFO_CIK_MACROTILE_MODE_ARRAY, NULL,
ws->info.cik_macrotile_mode_array)) {
- fprintf(stderr, "radeon: Kernel 3.13 is required for CIK support.\n");
+ fprintf(stderr, "radeon: Kernel 3.13 is required for Sea Islands support.\n");
return false;
}
}
- if (ws->info.chip_class >= SI) {
+ if (ws->info.chip_class >= GFX6) {
if (!radeon_get_drm_value(ws->fd, RADEON_INFO_SI_TILE_MODE_ARRAY, NULL,
ws->info.si_tile_mode_array)) {
- fprintf(stderr, "radeon: Kernel 3.10 is required for SI support.\n");
+ fprintf(stderr, "radeon: Kernel 3.10 is required for Southern Islands support.\n");
return false;
}
@@ -562,5 +562,5 @@ static bool do_winsys_init(struct radeon_drm_winsys *ws)
* accel_working2 with value 3 indicates the new firmware.
*/
- ws->info.gfx_ib_pad_with_type2 = ws->info.chip_class <= SI ||
+ ws->info.gfx_ib_pad_with_type2 = ws->info.chip_class <= GFX6 ||
(ws->info.family == CHIP_HAWAII &&
ws->accel_working2 < 3);
@@ -568,6 +568,6 @@ static bool do_winsys_init(struct radeon_drm_winsys *ws)
ws->info.ib_start_alignment = 4096;
ws->info.kernel_flushes_hdp_before_ib = ws->info.drm_minor >= 40;
- /* HTILE is broken with 1D tiling on old kernels and CIK. */
- ws->info.htile_cmask_support_1d_tiling = ws->info.chip_class != CIK ||
+ /* HTILE is broken with 1D tiling on old kernels and GFX7. */
+ ws->info.htile_cmask_support_1d_tiling = ws->info.chip_class != GFX7 ||
ws->info.drm_minor >= 38;
ws->info.si_TA_CS_BC_BASE_ADDR_allowed = ws->info.drm_minor >= 48;
@@ -579,13 +579,13 @@ static bool do_winsys_init(struct radeon_drm_winsys *ws)
/* Old kernels disallowed register writes via COPY_DATA
* that are used for indirect compute dispatches. */
- ws->info.has_indirect_compute_dispatch = ws->info.chip_class == CIK ||
- (ws->info.chip_class == SI &&
+ ws->info.has_indirect_compute_dispatch = ws->info.chip_class == GFX7 ||
+ (ws->info.chip_class == GFX6 &&
ws->info.drm_minor >= 45);
- /* SI doesn't support unaligned loads. */
- ws->info.has_unaligned_shader_loads = ws->info.chip_class == CIK &&
+ /* GFX6 doesn't support unaligned loads. */
+ ws->info.has_unaligned_shader_loads = ws->info.chip_class == GFX7 &&
ws->info.drm_minor >= 50;
ws->info.has_sparse_vm_mappings = false;
- /* 2D tiling on CIK is supported since DRM 2.35.0 */
- ws->info.has_2d_tiling = ws->info.chip_class <= SI || ws->info.drm_minor >= 35;
+ /* 2D tiling on GFX7 is supported since DRM 2.35.0 */
+ ws->info.has_2d_tiling = ws->info.chip_class <= GFX6 || ws->info.drm_minor >= 35;
ws->info.has_read_registers_query = ws->info.drm_minor >= 42;
ws->info.max_alignment = 1024*1024;
diff --git a/src/mesa/state_tracker/st_draw.c b/src/mesa/state_tracker/st_draw.c
index 266695f0c03..a5e93186bf5 100644
--- a/src/mesa/state_tracker/st_draw.c
+++ b/src/mesa/state_tracker/st_draw.c
@@ -90,5 +90,5 @@ setup_primitive_restart(struct gl_context *ctx, struct pipe_draw_info *info)
/* Enable primitive restart only when the restart index can have an
- * effect. This is required for correctness in radeonsi VI support.
+ * effect. This is required for correctness in radeonsi GFX8 support.
* Other hardware may also benefit from taking a faster, non-restart path
* when possible.
--
2.17.1
More information about the mesa-dev
mailing list