[Mesa-dev] [PATCH 12/18] ac/radeonsi: add tcs_rel_ids to the abi
Timothy Arceri
tarceri at itsqueeze.com
Mon Dec 11 02:43:22 UTC 2017
---
src/amd/common/ac_nir_to_llvm.c | 15 +++++++--------
src/amd/common/ac_shader_abi.h | 1 +
src/gallium/drivers/radeonsi/si_shader.c | 19 ++++++++++---------
src/gallium/drivers/radeonsi/si_shader_internal.h | 1 -
4 files changed, 18 insertions(+), 18 deletions(-)
diff --git a/src/amd/common/ac_nir_to_llvm.c b/src/amd/common/ac_nir_to_llvm.c
index 9641b43421d..ce91d403884 100644
--- a/src/amd/common/ac_nir_to_llvm.c
+++ b/src/amd/common/ac_nir_to_llvm.c
@@ -103,21 +103,20 @@ struct nir_to_llvm_context {
LLVMValueRef ls_out_layout;
LLVMValueRef es2gs_offset;
LLVMValueRef tcs_offchip_layout;
LLVMValueRef tcs_out_offsets;
LLVMValueRef tcs_out_layout;
LLVMValueRef tcs_in_layout;
LLVMValueRef oc_lds;
LLVMValueRef merged_wave_info;
LLVMValueRef tess_factor_offset;
- LLVMValueRef tcs_rel_ids;
LLVMValueRef tes_rel_patch_id;
LLVMValueRef tes_u;
LLVMValueRef tes_v;
LLVMValueRef gsvs_ring_stride;
LLVMValueRef gsvs_num_entries;
LLVMValueRef gs2vs_offset;
LLVMValueRef gs_wave_id;
LLVMValueRef gs_vtx_offset[6];
@@ -412,21 +411,21 @@ static LLVMValueRef unpack_param(struct ac_llvm_context *ctx,
value = LLVMBuildAnd(ctx->builder, value,
LLVMConstInt(ctx->i32, mask, false), "");
}
return value;
}
static LLVMValueRef get_rel_patch_id(struct nir_to_llvm_context *ctx)
{
switch (ctx->stage) {
case MESA_SHADER_TESS_CTRL:
- return unpack_param(&ctx->ac, ctx->tcs_rel_ids, 0, 8);
+ return unpack_param(&ctx->ac, ctx->abi.tcs_rel_ids, 0, 8);
case MESA_SHADER_TESS_EVAL:
return ctx->tes_rel_patch_id;
break;
default:
unreachable("Illegal stage");
}
}
/* Tessellation shaders pass outputs to the next shader using LDS.
*
@@ -781,37 +780,37 @@ static void create_function(struct nir_to_llvm_context *ctx,
add_user_sgpr_argument(&args, ctx->ac.i32, &ctx->ls_out_layout); // ls out layout
add_user_sgpr_argument(&args, ctx->ac.i32, &ctx->tcs_offchip_layout); // tcs offchip layout
add_user_sgpr_argument(&args, ctx->ac.i32, &ctx->tcs_out_offsets); // tcs out offsets
add_user_sgpr_argument(&args, ctx->ac.i32, &ctx->tcs_out_layout); // tcs out layout
add_user_sgpr_argument(&args, ctx->ac.i32, &ctx->tcs_in_layout); // tcs in layout
if (ctx->shader_info->info.needs_multiview_view_index)
add_user_sgpr_argument(&args, ctx->ac.i32, &ctx->view_index);
add_vgpr_argument(&args, ctx->ac.i32, &ctx->abi.tcs_patch_id); // patch id
- add_vgpr_argument(&args, ctx->ac.i32, &ctx->tcs_rel_ids); // rel ids;
+ add_vgpr_argument(&args, ctx->ac.i32, &ctx->abi.tcs_rel_ids); // rel ids;
add_vgpr_argument(&args, ctx->ac.i32, &ctx->abi.vertex_id); // vertex id
add_vgpr_argument(&args, ctx->ac.i32, &ctx->rel_auto_id); // rel auto id
add_vgpr_argument(&args, ctx->ac.i32, &ctx->vs_prim_id); // vs prim id
add_vgpr_argument(&args, ctx->ac.i32, &ctx->abi.instance_id); // instance id
} else {
radv_define_common_user_sgprs_phase1(ctx, stage, has_previous_stage, previous_stage, &user_sgpr_info, &args, &desc_sets);
add_user_sgpr_argument(&args, ctx->ac.i32, &ctx->tcs_offchip_layout); // tcs offchip layout
add_user_sgpr_argument(&args, ctx->ac.i32, &ctx->tcs_out_offsets); // tcs out offsets
add_user_sgpr_argument(&args, ctx->ac.i32, &ctx->tcs_out_layout); // tcs out layout
add_user_sgpr_argument(&args, ctx->ac.i32, &ctx->tcs_in_layout); // tcs in layout
if (ctx->shader_info->info.needs_multiview_view_index)
add_user_sgpr_argument(&args, ctx->ac.i32, &ctx->view_index);
add_sgpr_argument(&args, ctx->ac.i32, &ctx->oc_lds); // param oc lds
add_sgpr_argument(&args, ctx->ac.i32, &ctx->tess_factor_offset); // tess factor offset
add_vgpr_argument(&args, ctx->ac.i32, &ctx->abi.tcs_patch_id); // patch id
- add_vgpr_argument(&args, ctx->ac.i32, &ctx->tcs_rel_ids); // rel ids;
+ add_vgpr_argument(&args, ctx->ac.i32, &ctx->abi.tcs_rel_ids); // rel ids;
}
break;
case MESA_SHADER_TESS_EVAL:
radv_define_common_user_sgprs_phase1(ctx, stage, has_previous_stage, previous_stage, &user_sgpr_info, &args, &desc_sets);
add_user_sgpr_argument(&args, ctx->ac.i32, &ctx->tcs_offchip_layout); // tcs offchip layout
if (ctx->shader_info->info.needs_multiview_view_index || (!ctx->options->key.tes.as_es && ctx->options->key.has_multiview_view_index))
add_user_sgpr_argument(&args, ctx->ac.i32, &ctx->view_index);
if (ctx->options->key.tes.as_es) {
add_sgpr_argument(&args, ctx->ac.i32, &ctx->oc_lds); // OC LDS
add_sgpr_argument(&args, ctx->ac.i32, NULL); //
@@ -4070,21 +4069,21 @@ static void visit_intrinsic(struct ac_nir_context *ctx,
result = ctx->abi->start_instance;
break;
case nir_intrinsic_load_draw_id:
result = ctx->abi->draw_id;
break;
case nir_intrinsic_load_view_index:
result = ctx->nctx->view_index ? ctx->nctx->view_index : ctx->ac.i32_0;
break;
case nir_intrinsic_load_invocation_id:
if (ctx->stage == MESA_SHADER_TESS_CTRL)
- result = unpack_param(&ctx->ac, ctx->nctx->tcs_rel_ids, 8, 5);
+ result = unpack_param(&ctx->ac, ctx->abi->tcs_rel_ids, 8, 5);
else
result = ctx->abi->gs_invocation_id;
break;
case nir_intrinsic_load_primitive_id:
if (ctx->stage == MESA_SHADER_GEOMETRY) {
if (ctx->nctx)
ctx->nctx->shader_info->gs.uses_prim_id = true;
result = ctx->abi->gs_prim_id;
} else if (ctx->stage == MESA_SHADER_TESS_CTRL) {
if (ctx->nctx)
@@ -6008,22 +6007,22 @@ ac_nir_build_endif(struct ac_build_if_state *ifthen)
/* Resume building code at end of the ifthen->merge_block */
LLVMPositionBuilderAtEnd(builder, ifthen->merge_block);
}
static void
write_tess_factors(struct nir_to_llvm_context *ctx)
{
unsigned stride, outer_comps, inner_comps;
struct ac_build_if_state if_ctx, inner_if_ctx;
- LLVMValueRef invocation_id = unpack_param(&ctx->ac, ctx->tcs_rel_ids, 8, 5);
- LLVMValueRef rel_patch_id = unpack_param(&ctx->ac, ctx->tcs_rel_ids, 0, 8);
+ LLVMValueRef invocation_id = unpack_param(&ctx->ac, ctx->abi.tcs_rel_ids, 8, 5);
+ LLVMValueRef rel_patch_id = unpack_param(&ctx->ac, ctx->abi.tcs_rel_ids, 0, 8);
unsigned tess_inner_index, tess_outer_index;
LLVMValueRef lds_base, lds_inner, lds_outer, byteoffset, buffer;
LLVMValueRef out[6], vec0, vec1, tf_base, inner[4], outer[4];
int i;
emit_barrier(ctx);
switch (ctx->options->key.tcs.primitive_mode) {
case GL_ISOLINES:
stride = 2;
outer_comps = 2;
@@ -6429,21 +6428,21 @@ ac_nir_get_max_workgroup_size(enum chip_class chip_class,
/* Fixup the HW not emitting the TCS regs if there are no HS threads. */
static void ac_nir_fixup_ls_hs_input_vgprs(struct nir_to_llvm_context *ctx)
{
LLVMValueRef count = ac_build_bfe(&ctx->ac, ctx->merged_wave_info,
LLVMConstInt(ctx->ac.i32, 8, false),
LLVMConstInt(ctx->ac.i32, 8, false), false);
LLVMValueRef hs_empty = LLVMBuildICmp(ctx->ac.builder, LLVMIntEQ, count,
LLVMConstInt(ctx->ac.i32, 0, false), "");
ctx->abi.instance_id = LLVMBuildSelect(ctx->ac.builder, hs_empty, ctx->rel_auto_id, ctx->abi.instance_id, "");
ctx->vs_prim_id = LLVMBuildSelect(ctx->ac.builder, hs_empty, ctx->abi.vertex_id, ctx->vs_prim_id, "");
- ctx->rel_auto_id = LLVMBuildSelect(ctx->ac.builder, hs_empty, ctx->tcs_rel_ids, ctx->rel_auto_id, "");
+ ctx->rel_auto_id = LLVMBuildSelect(ctx->ac.builder, hs_empty, ctx->abi.tcs_rel_ids, ctx->rel_auto_id, "");
ctx->abi.vertex_id = LLVMBuildSelect(ctx->ac.builder, hs_empty, ctx->abi.tcs_patch_id, ctx->abi.vertex_id, "");
}
static void prepare_gs_input_vgprs(struct nir_to_llvm_context *ctx)
{
for(int i = 5; i >= 0; --i) {
ctx->gs_vtx_offset[i] = ac_build_bfe(&ctx->ac, ctx->gs_vtx_offset[i & ~1],
LLVMConstInt(ctx->ac.i32, (i & 1) * 16, false),
LLVMConstInt(ctx->ac.i32, 16, false), false);
}
diff --git a/src/amd/common/ac_shader_abi.h b/src/amd/common/ac_shader_abi.h
index 2aadc20d403..827617c8e97 100644
--- a/src/amd/common/ac_shader_abi.h
+++ b/src/amd/common/ac_shader_abi.h
@@ -36,20 +36,21 @@ enum ac_descriptor_type {
/* Document the shader ABI during compilation. This is what allows radeonsi and
* radv to share a compiler backend.
*/
struct ac_shader_abi {
LLVMValueRef base_vertex;
LLVMValueRef start_instance;
LLVMValueRef draw_id;
LLVMValueRef vertex_id;
LLVMValueRef instance_id;
LLVMValueRef tcs_patch_id;
+ LLVMValueRef tcs_rel_ids;
LLVMValueRef tes_patch_id;
LLVMValueRef gs_prim_id;
LLVMValueRef gs_invocation_id;
LLVMValueRef frag_pos[4];
LLVMValueRef front_face;
LLVMValueRef ancillary;
LLVMValueRef sample_coverage;
/* For VS and PS: pre-loaded shader inputs.
*
diff --git a/src/gallium/drivers/radeonsi/si_shader.c b/src/gallium/drivers/radeonsi/si_shader.c
index dcde83cb9c2..66d38415602 100644
--- a/src/gallium/drivers/radeonsi/si_shader.c
+++ b/src/gallium/drivers/radeonsi/si_shader.c
@@ -266,21 +266,21 @@ static LLVMValueRef unpack_param(struct si_shader_context *ctx,
{
LLVMValueRef value = LLVMGetParam(ctx->main_fn, param);
return unpack_llvm_param(ctx, value, rshift, bitwidth);
}
static LLVMValueRef get_rel_patch_id(struct si_shader_context *ctx)
{
switch (ctx->type) {
case PIPE_SHADER_TESS_CTRL:
- return unpack_param(ctx, ctx->param_tcs_rel_ids, 0, 8);
+ return unpack_llvm_param(ctx, ctx->abi.tcs_rel_ids, 0, 8);
case PIPE_SHADER_TESS_EVAL:
return LLVMGetParam(ctx->main_fn,
ctx->param_tes_rel_patch_id);
default:
assert(0);
return NULL;
}
}
@@ -1948,21 +1948,21 @@ void si_load_system_value(struct si_shader_context *ctx,
case TGSI_SEMANTIC_BASEINSTANCE:
value = ctx->abi.start_instance;
break;
case TGSI_SEMANTIC_DRAWID:
value = ctx->abi.draw_id;
break;
case TGSI_SEMANTIC_INVOCATIONID:
if (ctx->type == PIPE_SHADER_TESS_CTRL)
- value = unpack_param(ctx, ctx->param_tcs_rel_ids, 8, 5);
+ value = unpack_llvm_param(ctx, ctx->abi.tcs_rel_ids, 8, 5);
else if (ctx->type == PIPE_SHADER_GEOMETRY)
value = ctx->abi.gs_invocation_id;
else
assert(!"INVOCATIONID not implemented");
break;
case TGSI_SEMANTIC_POSITION:
{
LLVMValueRef pos[4] = {
LLVMGetParam(ctx->main_fn, SI_PARAM_POS_X_FLOAT),
@@ -2997,21 +2997,21 @@ static void si_llvm_export_vs(struct si_shader_context *ctx,
* Forward all outputs from the vertex shader to the TES. This is only used
* for the fixed function TCS.
*/
static void si_copy_tcs_inputs(struct lp_build_tgsi_context *bld_base)
{
struct si_shader_context *ctx = si_shader_context(bld_base);
LLVMValueRef invocation_id, buffer, buffer_offset;
LLVMValueRef lds_vertex_stride, lds_vertex_offset, lds_base;
uint64_t inputs;
- invocation_id = unpack_param(ctx, ctx->param_tcs_rel_ids, 8, 5);
+ invocation_id = unpack_llvm_param(ctx, ctx->abi.tcs_rel_ids, 8, 5);
buffer = desc_from_addr_base64k(ctx, ctx->param_tcs_offchip_addr_base64k);
buffer_offset = LLVMGetParam(ctx->main_fn, ctx->param_tcs_offchip_offset);
lds_vertex_stride = get_tcs_in_vertex_dw_stride(ctx);
lds_vertex_offset = LLVMBuildMul(ctx->ac.builder, invocation_id,
lds_vertex_stride, "");
lds_base = get_tcs_in_current_patch_offset(ctx);
lds_base = LLVMBuildAdd(ctx->ac.builder, lds_base, lds_vertex_offset, "");
inputs = ctx->shader->key.mono.u.ff_tcs_inputs_to_copy;
@@ -3250,21 +3250,21 @@ si_insert_input_ptr_as_2xi32(struct si_shader_context *ctx, LLVMValueRef ret,
/* This only writes the tessellation factor levels. */
static void si_llvm_emit_tcs_epilogue(struct lp_build_tgsi_context *bld_base)
{
struct si_shader_context *ctx = si_shader_context(bld_base);
LLVMBuilderRef builder = ctx->ac.builder;
LLVMValueRef rel_patch_id, invocation_id, tf_lds_offset;
si_copy_tcs_inputs(bld_base);
rel_patch_id = get_rel_patch_id(ctx);
- invocation_id = unpack_param(ctx, ctx->param_tcs_rel_ids, 8, 5);
+ invocation_id = unpack_llvm_param(ctx, ctx->abi.tcs_rel_ids, 8, 5);
tf_lds_offset = get_tcs_out_current_patch_data_offset(ctx);
if (ctx->screen->info.chip_class >= GFX9) {
LLVMBasicBlockRef blocks[2] = {
LLVMGetInsertBlock(builder),
ctx->merged_wrap_if_state.entry_block
};
LLVMValueRef values[2];
lp_build_endif(&ctx->merged_wrap_if_state);
@@ -3311,21 +3311,21 @@ static void si_llvm_emit_tcs_epilogue(struct lp_build_tgsi_context *bld_base)
GFX6_TCS_NUM_USER_SGPR + 1);
vgpr = GFX6_TCS_NUM_USER_SGPR + 2;
}
/* VGPRs */
rel_patch_id = ac_to_float(&ctx->ac, rel_patch_id);
invocation_id = ac_to_float(&ctx->ac, invocation_id);
tf_lds_offset = ac_to_float(&ctx->ac, tf_lds_offset);
/* Leave a hole corresponding to the two input VGPRs. This ensures that
- * the invocation_id output does not alias the param_tcs_rel_ids input,
+ * the invocation_id output does not alias the tcs_rel_ids input,
* which saves a V_MOV on gfx9.
*/
vgpr += 2;
ret = LLVMBuildInsertValue(builder, ret, rel_patch_id, vgpr++, "");
ret = LLVMBuildInsertValue(builder, ret, invocation_id, vgpr++, "");
if (ctx->shader->selector->tcs_info.tessfactors_are_def_in_all_invocs) {
vgpr++; /* skip the tess factor LDS offset */
for (unsigned i = 0; i < 6; i++) {
@@ -3372,22 +3372,23 @@ static void si_set_ls_return_value_for_tcs(struct si_shader_context *ctx)
unsigned desc_param = ctx->param_tcs_factor_addr_base64k + 2;
ret = si_insert_input_ptr_as_2xi32(ctx, ret, desc_param,
8 + GFX9_SGPR_TCS_CONST_AND_SHADER_BUFFERS);
ret = si_insert_input_ptr_as_2xi32(ctx, ret, desc_param + 1,
8 + GFX9_SGPR_TCS_SAMPLERS_AND_IMAGES);
unsigned vgpr = 8 + GFX9_TCS_NUM_USER_SGPR;
ret = LLVMBuildInsertValue(ctx->ac.builder, ret,
ac_to_float(&ctx->ac, ctx->abi.tcs_patch_id),
vgpr++, "");
- ret = si_insert_input_ret_float(ctx, ret,
- ctx->param_tcs_rel_ids, vgpr++);
+ ret = LLVMBuildInsertValue(ctx->ac.builder, ret,
+ ac_to_float(&ctx->ac, ctx->abi.tcs_rel_ids),
+ vgpr++, "");
ctx->return_value = ret;
}
/* Pass GS inputs from ES to GS on GFX9. */
static void si_set_es_return_value_for_gs(struct si_shader_context *ctx)
{
LLVMValueRef ret = ctx->return_value;
ret = si_insert_input_ret(ctx, ret, ctx->param_gs2vs_offset, 2);
ret = si_insert_input_ret(ctx, ret, ctx->param_merged_wave_info, 3);
@@ -4756,21 +4757,21 @@ static void create_function(struct si_shader_context *ctx)
ctx->param_tcs_out_lds_offsets = add_arg(&fninfo, ARG_SGPR, ctx->i32);
ctx->param_tcs_out_lds_layout = add_arg(&fninfo, ARG_SGPR, ctx->i32);
ctx->param_vs_state_bits = add_arg(&fninfo, ARG_SGPR, ctx->i32);
ctx->param_tcs_offchip_addr_base64k = add_arg(&fninfo, ARG_SGPR, ctx->i32);
ctx->param_tcs_factor_addr_base64k = add_arg(&fninfo, ARG_SGPR, ctx->i32);
ctx->param_tcs_offchip_offset = add_arg(&fninfo, ARG_SGPR, ctx->i32);
ctx->param_tcs_factor_offset = add_arg(&fninfo, ARG_SGPR, ctx->i32);
/* VGPRs */
add_arg_assign(&fninfo, ARG_VGPR, ctx->i32, &ctx->abi.tcs_patch_id);
- ctx->param_tcs_rel_ids = add_arg(&fninfo, ARG_VGPR, ctx->i32);
+ add_arg_assign(&fninfo, ARG_VGPR, ctx->i32, &ctx->abi.tcs_rel_ids);
/* param_tcs_offchip_offset and param_tcs_factor_offset are
* placed after the user SGPRs.
*/
for (i = 0; i < GFX6_TCS_NUM_USER_SGPR + 2; i++)
returns[num_returns++] = ctx->i32; /* SGPRs */
for (i = 0; i < 11; i++)
returns[num_returns++] = ctx->f32; /* VGPRs */
break;
@@ -4795,21 +4796,21 @@ static void create_function(struct si_shader_context *ctx)
ctx->param_tcs_out_lds_layout = add_arg(&fninfo, ARG_SGPR, ctx->i32);
ctx->param_tcs_offchip_addr_base64k = add_arg(&fninfo, ARG_SGPR, ctx->i32);
ctx->param_tcs_factor_addr_base64k = add_arg(&fninfo, ARG_SGPR, ctx->i32);
add_arg(&fninfo, ARG_SGPR, ctx->i32); /* unused */
declare_per_stage_desc_pointers(ctx, &fninfo,
ctx->type == PIPE_SHADER_TESS_CTRL);
/* VGPRs (first TCS, then VS) */
add_arg_assign(&fninfo, ARG_VGPR, ctx->i32, &ctx->abi.tcs_patch_id);
- ctx->param_tcs_rel_ids = add_arg(&fninfo, ARG_VGPR, ctx->i32);
+ add_arg_assign(&fninfo, ARG_VGPR, ctx->i32, &ctx->abi.tcs_rel_ids);
if (ctx->type == PIPE_SHADER_VERTEX) {
declare_vs_input_vgprs(ctx, &fninfo,
&num_prolog_vgprs);
/* LS return values are inputs to the TCS main shader part. */
for (i = 0; i < 8 + GFX9_TCS_NUM_USER_SGPR; i++)
returns[num_returns++] = ctx->i32; /* SGPRs */
for (i = 0; i < 2; i++)
returns[num_returns++] = ctx->f32; /* VGPRs */
diff --git a/src/gallium/drivers/radeonsi/si_shader_internal.h b/src/gallium/drivers/radeonsi/si_shader_internal.h
index 5c0d22dcd7f..33c6b0a26b9 100644
--- a/src/gallium/drivers/radeonsi/si_shader_internal.h
+++ b/src/gallium/drivers/radeonsi/si_shader_internal.h
@@ -162,21 +162,20 @@ struct si_shader_context {
/* Layout of TCS outputs / TES inputs:
* [0:12] = stride between output patches in DW, num_outputs * num_vertices * 4
* max = 32*32*4 + 32*4
* [26:31] = gl_PatchVerticesIn, max = 32
*/
int param_tcs_out_lds_layout;
int param_tcs_offchip_addr_base64k;
int param_tcs_factor_addr_base64k;
int param_tcs_offchip_offset;
int param_tcs_factor_offset;
- int param_tcs_rel_ids;
/* API TES */
int param_tes_u;
int param_tes_v;
int param_tes_rel_patch_id;
/* HW ES */
int param_es2gs_offset;
/* API GS */
int param_gs2vs_offset;
int param_gs_wave_id; /* GFX6 */
--
2.14.3
More information about the mesa-dev
mailing list