From: Marek Olšák <marek.ol...@amd.com>
They will vary with merged stages.
---
src/gallium/drivers/radeonsi/si_shader.c | 254 +++++++++++-----------
src/gallium/drivers/radeonsi/si_shader.h | 60 +----
src/gallium/drivers/radeonsi/si_shader_internal.h | 69 +++++-
3 files changed, 190 insertions(+), 193 deletions(-)
diff --git a/src/gallium/drivers/radeonsi/si_shader.c
b/src/gallium/drivers/radeonsi/si_shader.c
index 4ea1633..fbeb265 100644
--- a/src/gallium/drivers/radeonsi/si_shader.c
+++ b/src/gallium/drivers/radeonsi/si_shader.c
@@ -182,21 +182,21 @@ static LLVMValueRef unpack_param(struct si_shader_context
*ctx,
LLVMConstInt(ctx->i32, mask, 0), "");
}
return value;
}
static LLVMValueRef get_rel_patch_id(struct si_shader_context *ctx)
{
switch (ctx->type) {
case PIPE_SHADER_TESS_CTRL:
- return unpack_param(ctx, SI_PARAM_REL_IDS, 0, 8);
+ return unpack_param(ctx, ctx->param_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;
}
}
@@ -218,52 +218,45 @@ static LLVMValueRef get_rel_patch_id(struct
si_shader_context *ctx)
* - TCS outputs for patch 2 = get_tcs_out_current_patch_offset (if
RelPatchID==2)
* - Per-patch TCS outputs for patch 2 =
get_tcs_out_current_patch_data_offset (if RelPatchID==2)
* - ...
*
* All three shaders VS(LS), TCS, TES share the same LDS space.
*/
static LLVMValueRef
get_tcs_in_patch_stride(struct si_shader_context *ctx)
{
- if (ctx->type == PIPE_SHADER_VERTEX)
- return unpack_param(ctx, SI_PARAM_VS_STATE_BITS, 8, 13);
- else if (ctx->type == PIPE_SHADER_TESS_CTRL)
- return unpack_param(ctx, SI_PARAM_TCS_IN_LAYOUT, 8, 13);
- else {
- assert(0);
- return NULL;
- }
+ return unpack_param(ctx, ctx->param_vs_state_bits, 8, 13);
}
static LLVMValueRef
get_tcs_out_patch_stride(struct si_shader_context *ctx)
{
- return unpack_param(ctx, SI_PARAM_TCS_OUT_LAYOUT, 0, 13);
+ return unpack_param(ctx, ctx->param_tcs_out_lds_layout, 0, 13);
}
static LLVMValueRef
get_tcs_out_patch0_offset(struct si_shader_context *ctx)
{
return lp_build_mul_imm(&ctx->bld_base.uint_bld,
unpack_param(ctx,
- SI_PARAM_TCS_OUT_OFFSETS,
+ ctx->param_tcs_out_lds_offsets,
0, 16),
4);
}
static LLVMValueRef
get_tcs_out_patch0_patch_data_offset(struct si_shader_context *ctx)
{
return lp_build_mul_imm(&ctx->bld_base.uint_bld,
unpack_param(ctx,
- SI_PARAM_TCS_OUT_OFFSETS,
+ ctx->param_tcs_out_lds_offsets,
16, 16),
4);
}
static LLVMValueRef
get_tcs_in_current_patch_offset(struct si_shader_context *ctx)
{
struct gallivm_state *gallivm = &ctx->gallivm;
LLVMValueRef patch_stride = get_tcs_in_patch_stride(ctx);
LLVMValueRef rel_patch_id = get_rel_patch_id(ctx);
@@ -346,21 +339,21 @@ static void declare_input_vs(
unsigned num_fetches;
unsigned fetch_stride;
LLVMValueRef t_list_ptr;
LLVMValueRef t_offset;
LLVMValueRef t_list;
LLVMValueRef vertex_index;
LLVMValueRef input[3];
/* Load the T list */
- t_list_ptr = LLVMGetParam(ctx->main_fn, SI_PARAM_VERTEX_BUFFERS);
+ t_list_ptr = LLVMGetParam(ctx->main_fn, ctx->param_vertex_buffers);
t_offset = LLVMConstInt(ctx->i32, input_index, 0);
t_list = ac_build_indexed_load_const(&ctx->ac, t_list_ptr, t_offset);
vertex_index = LLVMGetParam(ctx->main_fn,
ctx->param_vertex_index0 +
input_index);
fix_fetch = ctx->shader->key.mono.vs_fix_fetch[input_index];
@@ -546,27 +539,27 @@ static LLVMValueRef get_primitive_id(struct
lp_build_tgsi_context *bld_base,
if (swizzle > 0)
return ctx->i32_0;
switch (ctx->type) {
case PIPE_SHADER_VERTEX:
return LLVMGetParam(ctx->main_fn,
ctx->param_vs_prim_id);
case PIPE_SHADER_TESS_CTRL:
return LLVMGetParam(ctx->main_fn,
- SI_PARAM_PATCH_ID);
+ ctx->param_tcs_patch_id);
case PIPE_SHADER_TESS_EVAL:
return LLVMGetParam(ctx->main_fn,
ctx->param_tes_patch_id);
case PIPE_SHADER_GEOMETRY:
return LLVMGetParam(ctx->main_fn,
- SI_PARAM_PRIMITIVE_ID);
+ ctx->param_gs_prim_id);
default:
assert(0);
return ctx->i32_0;
}
}
/**
* Return the value of tgsi_ind_register for indexing.
* This is the indirect index with the constant offset added to it.
*/
@@ -710,22 +703,22 @@ static LLVMValueRef get_dw_address(struct
si_shader_context *ctx,
*/
static LLVMValueRef get_tcs_tes_buffer_address(struct si_shader_context *ctx,
LLVMValueRef rel_patch_id,
LLVMValueRef vertex_index,
LLVMValueRef param_index)
{
struct gallivm_state *gallivm = &ctx->gallivm;
LLVMValueRef base_addr, vertices_per_patch, num_patches, total_vertices;
LLVMValueRef param_stride, constant16;
- vertices_per_patch = unpack_param(ctx, SI_PARAM_TCS_OFFCHIP_LAYOUT, 9,
6);
- num_patches = unpack_param(ctx, SI_PARAM_TCS_OFFCHIP_LAYOUT, 0, 9);
+ vertices_per_patch = unpack_param(ctx, ctx->param_tcs_offchip_layout,
9, 6);
+ num_patches = unpack_param(ctx, ctx->param_tcs_offchip_layout, 0, 9);
total_vertices = LLVMBuildMul(gallivm->builder, vertices_per_patch,
num_patches, "");
constant16 = LLVMConstInt(ctx->i32, 16, 0);
if (vertex_index) {
base_addr = LLVMBuildMul(gallivm->builder, rel_patch_id,
vertices_per_patch, "");
base_addr = LLVMBuildAdd(gallivm->builder, base_addr,
vertex_index, "");
@@ -737,21 +730,21 @@ static LLVMValueRef get_tcs_tes_buffer_address(struct
si_shader_context *ctx,
}
base_addr = LLVMBuildAdd(gallivm->builder, base_addr,
LLVMBuildMul(gallivm->builder, param_index,
param_stride, ""), "");
base_addr = LLVMBuildMul(gallivm->builder, base_addr, constant16, "");
if (!vertex_index) {
LLVMValueRef patch_data_offset =
- unpack_param(ctx, SI_PARAM_TCS_OFFCHIP_LAYOUT, 16,
16);
+ unpack_param(ctx, ctx->param_tcs_offchip_layout, 16,
16);
base_addr = LLVMBuildAdd(gallivm->builder, base_addr,
patch_data_offset, "");
}
return base_addr;
}
static LLVMValueRef get_tcs_tes_buffer_address_from_reg(
struct si_shader_context *ctx,
const struct tgsi_full_dst_register
*dst,
@@ -915,61 +908,61 @@ static void lds_store(struct lp_build_tgsi_context
*bld_base,
}
static LLVMValueRef fetch_input_tcs(
struct lp_build_tgsi_context *bld_base,
const struct tgsi_full_src_register *reg,
enum tgsi_opcode_type type, unsigned swizzle)
{
struct si_shader_context *ctx = si_shader_context(bld_base);
LLVMValueRef dw_addr, stride;
- stride = unpack_param(ctx, SI_PARAM_TCS_IN_LAYOUT, 24, 8);
+ stride = unpack_param(ctx, ctx->param_vs_state_bits, 24, 8);
dw_addr = get_tcs_in_current_patch_offset(ctx);
dw_addr = get_dw_address(ctx, NULL, reg, stride, dw_addr);
return lds_load(bld_base, type, swizzle, dw_addr);
}
static LLVMValueRef fetch_output_tcs(
struct lp_build_tgsi_context *bld_base,
const struct tgsi_full_src_register *reg,
enum tgsi_opcode_type type, unsigned swizzle)
{
struct si_shader_context *ctx = si_shader_context(bld_base);
LLVMValueRef dw_addr, stride;
if (reg->Register.Dimension) {
- stride = unpack_param(ctx, SI_PARAM_TCS_OUT_LAYOUT, 13, 8);
+ stride = unpack_param(ctx, ctx->param_tcs_out_lds_layout, 13,
8);
dw_addr = get_tcs_out_current_patch_offset(ctx);
dw_addr = get_dw_address(ctx, NULL, reg, stride, dw_addr);
} else {
dw_addr = get_tcs_out_current_patch_data_offset(ctx);
dw_addr = get_dw_address(ctx, NULL, reg, NULL, dw_addr);
}
return lds_load(bld_base, type, swizzle, dw_addr);
}
static LLVMValueRef fetch_input_tes(
struct lp_build_tgsi_context *bld_base,
const struct tgsi_full_src_register *reg,
enum tgsi_opcode_type type, unsigned swizzle)
{
struct si_shader_context *ctx = si_shader_context(bld_base);
LLVMValueRef rw_buffers, buffer, base, addr;
rw_buffers = LLVMGetParam(ctx->main_fn,
- SI_PARAM_RW_BUFFERS);
+ ctx->param_rw_buffers);
buffer = ac_build_indexed_load_const(&ctx->ac, rw_buffers,
LLVMConstInt(ctx->i32, SI_HS_RING_TESS_OFFCHIP, 0));
- base = LLVMGetParam(ctx->main_fn, ctx->param_oc_lds);
+ base = LLVMGetParam(ctx->main_fn, ctx->param_tcs_offchip_offset);
addr = get_tcs_tes_buffer_address_from_reg(ctx, NULL, reg);
return buffer_load(bld_base, type, swizzle, buffer, base, addr, true);
}
static void store_output_tcs(struct lp_build_tgsi_context *bld_base,
const struct tgsi_full_instruction *inst,
const struct tgsi_opcode_info *info,
LLVMValueRef dst[4])
{
@@ -987,21 +980,21 @@ static void store_output_tcs(struct lp_build_tgsi_context
*bld_base,
/* Only handle per-patch and per-vertex outputs here.
* Vectors will be lowered to scalars and this function will be called
again.
*/
if (reg->Register.File != TGSI_FILE_OUTPUT ||
(dst[0] && LLVMGetTypeKind(LLVMTypeOf(dst[0])) ==
LLVMVectorTypeKind)) {
si_llvm_emit_store(bld_base, inst, info, dst);
return;
}
if (reg->Register.Dimension) {
- stride = unpack_param(ctx, SI_PARAM_TCS_OUT_LAYOUT, 13, 8);
+ stride = unpack_param(ctx, ctx->param_tcs_out_lds_layout, 13,
8);
dw_addr = get_tcs_out_current_patch_offset(ctx);
dw_addr = get_dw_address(ctx, reg, NULL, stride, dw_addr);
skip_lds_store = !sh_info->reads_pervertex_outputs;
} else {
dw_addr = get_tcs_out_current_patch_data_offset(ctx);
dw_addr = get_dw_address(ctx, reg, NULL, NULL, dw_addr);
skip_lds_store = !sh_info->reads_perpatch_outputs;
if (!reg->Register.Indirect) {
int name =
sh_info->output_semantic_name[reg->Register.Index];
@@ -1009,25 +1002,25 @@ static void store_output_tcs(struct
lp_build_tgsi_context *bld_base,
/* Always write tess factors into LDS for the TCS
epilog. */
if (name == TGSI_SEMANTIC_TESSINNER ||
name == TGSI_SEMANTIC_TESSOUTER) {
skip_lds_store = false;
is_tess_factor = true;
}
}
}
rw_buffers = LLVMGetParam(ctx->main_fn,
- SI_PARAM_RW_BUFFERS);
+ ctx->param_rw_buffers);
buffer = ac_build_indexed_load_const(&ctx->ac, rw_buffers,
LLVMConstInt(ctx->i32, SI_HS_RING_TESS_OFFCHIP, 0));
- base = LLVMGetParam(ctx->main_fn, ctx->param_oc_lds);
+ base = LLVMGetParam(ctx->main_fn, ctx->param_tcs_offchip_offset);
buf_addr = get_tcs_tes_buffer_address_from_reg(ctx, reg, NULL);
TGSI_FOR_EACH_DST0_ENABLED_CHANNEL(inst, chan_index) {
LLVMValueRef value = dst[chan_index];
if (inst->Instruction.Saturate)
value = ac_build_clamp(&ctx->ac, value);
/* Skip LDS stores if there is no LDS read of this output. */
@@ -1082,24 +1075,24 @@ static LLVMValueRef fetch_input_gs(
for (chan = 0; chan < TGSI_NUM_CHANNELS; chan++) {
values[chan] = fetch_input_gs(bld_base, reg, type,
chan);
}
return lp_build_gather_values(gallivm, values,
TGSI_NUM_CHANNELS);
}
/* Get the vertex offset parameter */
vtx_offset_param = reg->Dimension.Index;
if (vtx_offset_param < 2) {
- vtx_offset_param += SI_PARAM_VTX0_OFFSET;
+ vtx_offset_param += ctx->param_gs_vtx0_offset;
} else {
assert(vtx_offset_param < 6);
- vtx_offset_param += SI_PARAM_VTX2_OFFSET - 2;
+ vtx_offset_param += ctx->param_gs_vtx2_offset - 2;
}
vtx_offset = lp_build_mul_imm(uint,
LLVMGetParam(ctx->main_fn,
vtx_offset_param),
4);
param = si_shader_io_get_unique_index(semantic_name, semantic_index);
soffset = LLVMConstInt(ctx->i32, (param * 4 + swizzle) * 256, 0);
value = ac_build_buffer_load(&ctx->ac, ctx->esgs_ring, 1, ctx->i32_0,
@@ -1344,21 +1337,21 @@ static LLVMValueRef buffer_load_const(struct
si_shader_context *ctx,
return lp_build_intrinsic(builder, "llvm.SI.load.const", ctx->f32,
args, 2,
LP_FUNC_ATTR_READNONE |
LP_FUNC_ATTR_LEGACY);
}
static LLVMValueRef load_sample_position(struct si_shader_context *ctx,
LLVMValueRef sample_id)
{
struct lp_build_context *uint_bld = &ctx->bld_base.uint_bld;
struct gallivm_state *gallivm = &ctx->gallivm;
LLVMBuilderRef builder = gallivm->builder;
- LLVMValueRef desc = LLVMGetParam(ctx->main_fn, SI_PARAM_RW_BUFFERS);
+ LLVMValueRef desc = LLVMGetParam(ctx->main_fn, ctx->param_rw_buffers);
LLVMValueRef buf_index = LLVMConstInt(ctx->i32,
SI_PS_CONST_SAMPLE_POSITIONS, 0);
LLVMValueRef resource = ac_build_indexed_load_const(&ctx->ac, desc,
buf_index);
/* offset = sample_id * 8 (8 = 2 floats containing samplepos.xy) */
LLVMValueRef offset0 = lp_build_mul_imm(uint_bld, sample_id, 8);
LLVMValueRef offset1 = LLVMBuildAdd(builder, offset0,
LLVMConstInt(ctx->i32, 4, 0), "");
LLVMValueRef pos[4] = {
buffer_load_const(ctx, resource, offset0),
buffer_load_const(ctx, resource, offset1),
@@ -1383,63 +1376,61 @@ static void declare_system_value(struct
si_shader_context *ctx,
case TGSI_SEMANTIC_INSTANCEID:
value = LLVMGetParam(ctx->main_fn,
ctx->param_instance_id);
break;
case TGSI_SEMANTIC_VERTEXID:
value = LLVMBuildAdd(gallivm->builder,
LLVMGetParam(ctx->main_fn,
ctx->param_vertex_id),
LLVMGetParam(ctx->main_fn,
- SI_PARAM_BASE_VERTEX), "");
+ ctx->param_base_vertex), "");
break;
case TGSI_SEMANTIC_VERTEXID_NOBASE:
/* Unused. Clarify the meaning in indexed vs. non-indexed
* draws if this is ever used again. */
assert(false);
break;
case TGSI_SEMANTIC_BASEVERTEX:
{
/* For non-indexed draws, the base vertex set by the driver
* (for direct draws) or the CP (for indirect draws) is the
* first vertex ID, but GLSL expects 0 to be returned.
*/
- LLVMValueRef vs_state = LLVMGetParam(ctx->main_fn,
SI_PARAM_VS_STATE_BITS);
+ LLVMValueRef vs_state = LLVMGetParam(ctx->main_fn,
ctx->param_vs_state_bits);
LLVMValueRef indexed;
indexed = LLVMBuildLShr(gallivm->builder, vs_state, ctx->i32_1,
"");
indexed = LLVMBuildTrunc(gallivm->builder, indexed, ctx->i1,
"");
value = LLVMBuildSelect(gallivm->builder, indexed,
- LLVMGetParam(ctx->main_fn,
SI_PARAM_BASE_VERTEX),
+ LLVMGetParam(ctx->main_fn,
ctx->param_base_vertex),
ctx->i32_0, "");
break;
}
case TGSI_SEMANTIC_BASEINSTANCE:
- value = LLVMGetParam(ctx->main_fn,
- SI_PARAM_START_INSTANCE);
+ value = LLVMGetParam(ctx->main_fn, ctx->param_start_instance);
break;
case TGSI_SEMANTIC_DRAWID:
- value = LLVMGetParam(ctx->main_fn,
- SI_PARAM_DRAWID);
+ value = LLVMGetParam(ctx->main_fn, ctx->param_draw_id);
break;
case TGSI_SEMANTIC_INVOCATIONID:
if (ctx->type == PIPE_SHADER_TESS_CTRL)
- value = unpack_param(ctx, SI_PARAM_REL_IDS, 8, 5);
+ value = unpack_param(ctx, ctx->param_tcs_rel_ids, 8, 5);
else if (ctx->type == PIPE_SHADER_GEOMETRY)
value = LLVMGetParam(ctx->main_fn,
- SI_PARAM_GS_INSTANCE_ID);
+ ctx->param_gs_instance_id);
else
assert(!"INVOCATIONID not implemented");
break;
case TGSI_SEMANTIC_POSITION:
{
LLVMValueRef pos[4] = {
LLVMGetParam(ctx->main_fn, SI_PARAM_POS_X_FLOAT),
LLVMGetParam(ctx->main_fn, SI_PARAM_POS_Y_FLOAT),
LLVMGetParam(ctx->main_fn, SI_PARAM_POS_Z_FLOAT),
@@ -1495,56 +1486,56 @@ static void declare_system_value(struct
si_shader_context *ctx,
PIPE_PRIM_TRIANGLES)
coord[2] = lp_build_sub(bld, bld->one,
lp_build_add(bld, coord[0],
coord[1]));
value = lp_build_gather_values(gallivm, coord, 4);
break;
}
case TGSI_SEMANTIC_VERTICESIN:
if (ctx->type == PIPE_SHADER_TESS_CTRL)
- value = unpack_param(ctx, SI_PARAM_TCS_OUT_LAYOUT, 26,
6);
+ value = unpack_param(ctx,
ctx->param_tcs_out_lds_layout, 26, 6);
else if (ctx->type == PIPE_SHADER_TESS_EVAL)
- value = unpack_param(ctx, SI_PARAM_TCS_OFFCHIP_LAYOUT,
9, 7);
+ value = unpack_param(ctx,
ctx->param_tcs_offchip_layout, 9, 7);
else
assert(!"invalid shader stage for
TGSI_SEMANTIC_VERTICESIN");
break;
case TGSI_SEMANTIC_TESSINNER:
case TGSI_SEMANTIC_TESSOUTER:
{
LLVMValueRef rw_buffers, buffer, base, addr;
int param = si_shader_io_get_unique_index(decl->Semantic.Name,
0);
rw_buffers = LLVMGetParam(ctx->main_fn,
- SI_PARAM_RW_BUFFERS);
+ ctx->param_rw_buffers);
buffer = ac_build_indexed_load_const(&ctx->ac, rw_buffers,
LLVMConstInt(ctx->i32, SI_HS_RING_TESS_OFFCHIP, 0));
- base = LLVMGetParam(ctx->main_fn, ctx->param_oc_lds);
+ base = LLVMGetParam(ctx->main_fn,
ctx->param_tcs_offchip_offset);
addr = get_tcs_tes_buffer_address(ctx, get_rel_patch_id(ctx),
NULL,
LLVMConstInt(ctx->i32, param, 0));
value = buffer_load(&ctx->bld_base, TGSI_TYPE_FLOAT,
~0, buffer, base, addr, true);
break;
}
case TGSI_SEMANTIC_DEFAULT_TESSOUTER_SI:
case TGSI_SEMANTIC_DEFAULT_TESSINNER_SI:
{
LLVMValueRef buf, slot, val[4];
int i, offset;
slot = LLVMConstInt(ctx->i32, SI_HS_CONST_DEFAULT_TESS_LEVELS,
0);
- buf = LLVMGetParam(ctx->main_fn, SI_PARAM_RW_BUFFERS);
+ buf = LLVMGetParam(ctx->main_fn, ctx->param_rw_buffers);
buf = ac_build_indexed_load_const(&ctx->ac, buf, slot);
offset = decl->Semantic.Name ==
TGSI_SEMANTIC_DEFAULT_TESSINNER_SI ? 4 : 0;
for (i = 0; i < 4; i++)
val[i] = buffer_load_const(ctx, buf,
LLVMConstInt(ctx->i32,
(offset + i) * 4, 0));
value = lp_build_gather_values(gallivm, val, 4);
break;
}
@@ -1667,21 +1658,21 @@ static void declare_compute_memory(struct
si_shader_context *ctx,
"compute_lds",
LOCAL_ADDR_SPACE);
LLVMSetAlignment(var, 4);
ctx->shared_memory = LLVMBuildBitCast(gallivm->builder, var, i8p, "");
}
static LLVMValueRef load_const_buffer_desc(struct si_shader_context *ctx, int
i)
{
LLVMValueRef list_ptr = LLVMGetParam(ctx->main_fn,
- SI_PARAM_CONST_BUFFERS);
+ ctx->param_const_buffers);
return ac_build_indexed_load_const(&ctx->ac, list_ptr,
LLVMConstInt(ctx->i32, i, 0));
}
static LLVMValueRef fetch_constant(
struct lp_build_tgsi_context *bld_base,
const struct tgsi_full_src_register *reg,
enum tgsi_opcode_type type,
unsigned swizzle)
@@ -1700,21 +1691,21 @@ static LLVMValueRef fetch_constant(
for (chan = 0; chan < TGSI_NUM_CHANNELS; ++chan)
values[chan] = fetch_constant(bld_base, reg, type,
chan);
return lp_build_gather_values(&ctx->gallivm, values, 4);
}
buf = reg->Register.Dimension ? reg->Dimension.Index : 0;
idx = reg->Register.Index * 4 + swizzle;
if (reg->Register.Dimension && reg->Dimension.Indirect) {
- LLVMValueRef ptr = LLVMGetParam(ctx->main_fn,
SI_PARAM_CONST_BUFFERS);
+ LLVMValueRef ptr = LLVMGetParam(ctx->main_fn,
ctx->param_const_buffers);
LLVMValueRef index;
index = get_bounded_indirect_index(ctx, ®->DimIndirect,
reg->Dimension.Index,
SI_NUM_CONST_BUFFERS);
bufp = ac_build_indexed_load_const(&ctx->ac, ptr, index);
} else
bufp = load_const_buffer_desc(ctx, buf);
if (reg->Register.Indirect) {
addr = ctx->addrs[ireg->Index][ireg->Swizzle];
@@ -2006,21 +1997,21 @@ static LLVMValueRef
si_scale_alpha_by_sample_mask(struct lp_build_tgsi_context *
static void si_llvm_emit_clipvertex(struct lp_build_tgsi_context *bld_base,
struct ac_export_args *pos, LLVMValueRef
*out_elts)
{
struct si_shader_context *ctx = si_shader_context(bld_base);
struct lp_build_context *base = &bld_base->base;
unsigned reg_index;
unsigned chan;
unsigned const_chan;
LLVMValueRef base_elt;
- LLVMValueRef ptr = LLVMGetParam(ctx->main_fn, SI_PARAM_RW_BUFFERS);
+ LLVMValueRef ptr = LLVMGetParam(ctx->main_fn, ctx->param_rw_buffers);
LLVMValueRef constbuf_index = LLVMConstInt(ctx->i32,
SI_VS_CONST_CLIP_PLANES, 0);
LLVMValueRef const_resource = ac_build_indexed_load_const(&ctx->ac,
ptr, constbuf_index);
for (reg_index = 0; reg_index < 2; reg_index ++) {
struct ac_export_args *args = &pos[2 + reg_index];
args->out[0] =
args->out[1] =
args->out[2] =
@@ -2162,21 +2153,21 @@ static void si_llvm_emit_streamout(struct
si_shader_context *ctx,
ctx->param_streamout_write_index);
/* Compute (streamout_write_index + thread_id). */
so_write_index = LLVMBuildAdd(builder, so_write_index, tid, "");
/* Load the descriptor and compute the write offset for each
* enabled buffer. */
LLVMValueRef so_write_offset[4] = {};
LLVMValueRef so_buffers[4];
LLVMValueRef buf_ptr = LLVMGetParam(ctx->main_fn,
- SI_PARAM_RW_BUFFERS);
+ ctx->param_rw_buffers);
for (i = 0; i < 4; i++) {
if (!so->stride[i])
continue;
LLVMValueRef offset = LLVMConstInt(ctx->i32,
SI_VS_STREAMOUT_BUF0
+ i, 0);
so_buffers[i] = ac_build_indexed_load_const(&ctx->ac,
buf_ptr, offset);
@@ -2405,29 +2396,29 @@ handle_semantic:
* 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);
struct gallivm_state *gallivm = &ctx->gallivm;
LLVMValueRef invocation_id, rw_buffers, buffer, buffer_offset;
LLVMValueRef lds_vertex_stride, lds_vertex_offset, lds_base;
uint64_t inputs;
- invocation_id = unpack_param(ctx, SI_PARAM_REL_IDS, 8, 5);
+ invocation_id = unpack_param(ctx, ctx->param_tcs_rel_ids, 8, 5);
- rw_buffers = LLVMGetParam(ctx->main_fn, SI_PARAM_RW_BUFFERS);
+ rw_buffers = LLVMGetParam(ctx->main_fn, ctx->param_rw_buffers);
buffer = ac_build_indexed_load_const(&ctx->ac, rw_buffers,
LLVMConstInt(ctx->i32, SI_HS_RING_TESS_OFFCHIP, 0));
- buffer_offset = LLVMGetParam(ctx->main_fn, ctx->param_oc_lds);
+ buffer_offset = LLVMGetParam(ctx->main_fn,
ctx->param_tcs_offchip_offset);
- lds_vertex_stride = unpack_param(ctx, SI_PARAM_TCS_IN_LAYOUT, 24, 8);
+ lds_vertex_stride = unpack_param(ctx, ctx->param_vs_state_bits, 24, 8);
lds_vertex_offset = LLVMBuildMul(gallivm->builder, invocation_id,
lds_vertex_stride, "");
lds_base = get_tcs_in_current_patch_offset(ctx);
lds_base = LLVMBuildAdd(gallivm->builder, lds_base, lds_vertex_offset,
"");
inputs = ctx->shader->key.mono.ff_tcs_inputs_to_copy;
while (inputs) {
unsigned i = u_bit_scan64(&inputs);
LLVMValueRef lds_ptr = LLVMBuildAdd(gallivm->builder, lds_base,
@@ -2533,27 +2524,27 @@ static void si_write_tess_factors(struct
lp_build_tgsi_context *bld_base,
/* Convert the outputs to vectors for stores. */
vec0 = lp_build_gather_values(gallivm, out, MIN2(stride, 4));
vec1 = NULL;
if (stride > 4)
vec1 = lp_build_gather_values(gallivm, out+4, stride - 4);
/* Get the buffer. */
rw_buffers = LLVMGetParam(ctx->main_fn,
- SI_PARAM_RW_BUFFERS);
+ ctx->param_rw_buffers);
buffer = ac_build_indexed_load_const(&ctx->ac, rw_buffers,
LLVMConstInt(ctx->i32, SI_HS_RING_TESS_FACTOR, 0));
/* Get the offset. */
tf_base = LLVMGetParam(ctx->main_fn,
- SI_PARAM_TESS_FACTOR_OFFSET);
+ ctx->param_tcs_factor_offset);
byteoffset = LLVMBuildMul(gallivm->builder, rel_patch_id,
LLVMConstInt(ctx->i32, 4 * stride, 0), "");
lp_build_if(&inner_if_ctx, gallivm,
LLVMBuildICmp(gallivm->builder, LLVMIntEQ,
rel_patch_id, ctx->i32_0, ""));
/* Store the dynamic HS control word. */
ac_build_buffer_store_dword(&ctx->ac, buffer,
LLVMConstInt(ctx->i32, 0x80000000, 0),
@@ -2572,21 +2563,21 @@ static void si_write_tess_factors(struct
lp_build_tgsi_context *bld_base,
20, 1, 0, true, false);
/* Store the tess factors into the offchip buffer if TES reads them. */
if (shader->key.part.tcs.epilog.tes_reads_tess_factors) {
LLVMValueRef buf, base, inner_vec, outer_vec, tf_outer_offset;
LLVMValueRef tf_inner_offset;
unsigned param_outer, param_inner;
buf = ac_build_indexed_load_const(&ctx->ac, rw_buffers,
LLVMConstInt(ctx->i32, SI_HS_RING_TESS_OFFCHIP,
0));
- base = LLVMGetParam(ctx->main_fn, ctx->param_oc_lds);
+ base = LLVMGetParam(ctx->main_fn,
ctx->param_tcs_offchip_offset);
param_outer = si_shader_io_get_unique_index(
TGSI_SEMANTIC_TESSOUTER, 0);
tf_outer_offset = get_tcs_tes_buffer_address(ctx, rel_patch_id,
NULL,
LLVMConstInt(ctx->i32, param_outer, 0));
outer_vec = lp_build_gather_values(gallivm, outer,
util_next_power_of_two(outer_comps));
ac_build_buffer_store_dword(&ctx->ac, buf, outer_vec,
@@ -2612,47 +2603,48 @@ static void si_write_tess_factors(struct
lp_build_tgsi_context *bld_base,
/* 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);
LLVMValueRef rel_patch_id, invocation_id, tf_lds_offset;
LLVMValueRef offchip_soffset, offchip_layout;
si_copy_tcs_inputs(bld_base);
rel_patch_id = get_rel_patch_id(ctx);
- invocation_id = unpack_param(ctx, SI_PARAM_REL_IDS, 8, 5);
+ invocation_id = unpack_param(ctx, ctx->param_tcs_rel_ids, 8, 5);
tf_lds_offset = get_tcs_out_current_patch_data_offset(ctx);
/* Return epilog parameters from this function. */
LLVMBuilderRef builder = ctx->gallivm.builder;
LLVMValueRef ret = ctx->return_value;
LLVMValueRef rw_buffers, rw0, rw1, tf_soffset;
unsigned vgpr;
/* RW_BUFFERS pointer */
rw_buffers = LLVMGetParam(ctx->main_fn,
- SI_PARAM_RW_BUFFERS);
+ ctx->param_rw_buffers);
rw_buffers = LLVMBuildPtrToInt(builder, rw_buffers, ctx->i64, "");
rw_buffers = LLVMBuildBitCast(builder, rw_buffers, ctx->v2i32, "");
rw0 = LLVMBuildExtractElement(builder, rw_buffers,
ctx->i32_0, "");
rw1 = LLVMBuildExtractElement(builder, rw_buffers,
ctx->i32_1, "");
ret = LLVMBuildInsertValue(builder, ret, rw0, 0, "");
ret = LLVMBuildInsertValue(builder, ret, rw1, 1, "");
/* Tess offchip and factor buffer soffset are after user SGPRs. */
offchip_layout = LLVMGetParam(ctx->main_fn,
- SI_PARAM_TCS_OFFCHIP_LAYOUT);
- offchip_soffset = LLVMGetParam(ctx->main_fn, ctx->param_oc_lds);
+ ctx->param_tcs_offchip_layout);
+ offchip_soffset = LLVMGetParam(ctx->main_fn,
+ ctx->param_tcs_offchip_offset);
tf_soffset = LLVMGetParam(ctx->main_fn,
- SI_PARAM_TESS_FACTOR_OFFSET);
+ ctx->param_tcs_factor_offset);
ret = LLVMBuildInsertValue(builder, ret, offchip_layout,
GFX6_SGPR_TCS_OFFCHIP_LAYOUT, "");
ret = LLVMBuildInsertValue(builder, ret, offchip_soffset,
GFX6_TCS_NUM_USER_SGPR, "");
ret = LLVMBuildInsertValue(builder, ret, tf_soffset,
GFX6_TCS_NUM_USER_SGPR + 1, "");
/* VGPRs */
rel_patch_id = bitcast(bld_base, TGSI_TYPE_FLOAT, rel_patch_id);
invocation_id = bitcast(bld_base, TGSI_TYPE_FLOAT, invocation_id);
@@ -2668,21 +2660,21 @@ static void si_llvm_emit_tcs_epilogue(struct
lp_build_tgsi_context *bld_base)
static void si_llvm_emit_ls_epilogue(struct lp_build_tgsi_context *bld_base)
{
struct si_shader_context *ctx = si_shader_context(bld_base);
struct si_shader *shader = ctx->shader;
struct tgsi_shader_info *info = &shader->selector->info;
struct gallivm_state *gallivm = &ctx->gallivm;
unsigned i, chan;
LLVMValueRef vertex_id = LLVMGetParam(ctx->main_fn,
ctx->param_rel_auto_id);
LLVMValueRef vertex_dw_stride =
- unpack_param(ctx, SI_PARAM_VS_STATE_BITS, 24, 8);
+ unpack_param(ctx, ctx->param_vs_state_bits, 24, 8);
LLVMValueRef base_dw_addr = LLVMBuildMul(gallivm->builder, vertex_id,
vertex_dw_stride, "");
/* Write outputs to LDS. The next shader (TCS aka HS) will read
* its inputs from it. */
for (i = 0; i < info->num_outputs; i++) {
LLVMValueRef *out_ptr = ctx->outputs[i];
unsigned name = info->output_semantic_name[i];
unsigned index = info->output_semantic_index[i];
@@ -2749,21 +2741,21 @@ static void si_llvm_emit_es_epilogue(struct
lp_build_tgsi_context *bld_base)
1, 1, true, true);
}
}
}
static void si_llvm_emit_gs_epilogue(struct lp_build_tgsi_context *bld_base)
{
struct si_shader_context *ctx = si_shader_context(bld_base);
ac_build_sendmsg(&ctx->ac, AC_SENDMSG_GS_OP_NOP | AC_SENDMSG_GS_DONE,
- LLVMGetParam(ctx->main_fn, SI_PARAM_GS_WAVE_ID));
+ LLVMGetParam(ctx->main_fn, ctx->param_gs_wave_id));
}
static void si_llvm_emit_vs_epilogue(struct lp_build_tgsi_context *bld_base)
{
struct si_shader_context *ctx = si_shader_context(bld_base);
struct gallivm_state *gallivm = &ctx->gallivm;
struct tgsi_shader_info *info = &ctx->shader->selector->info;
struct si_shader_output_values *outputs = NULL;
int i,j;
@@ -2784,21 +2776,21 @@ static void si_llvm_emit_vs_epilogue(struct
lp_build_tgsi_context *bld_base)
for (i = 0; i < info->num_outputs; i++) {
if (info->output_semantic_name[i] !=
TGSI_SEMANTIC_COLOR &&
info->output_semantic_name[i] !=
TGSI_SEMANTIC_BCOLOR)
continue;
/* We've found a color. */
if (!cond) {
/* The state is in the first bit of the user
SGPR. */
cond = LLVMGetParam(ctx->main_fn,
- SI_PARAM_VS_STATE_BITS);
+ ctx->param_vs_state_bits);
cond = LLVMBuildTrunc(gallivm->builder, cond,
ctx->i1, "");
lp_build_if(&if_ctx, gallivm, cond);
}
for (j = 0; j < 4; j++) {
addr = ctx->outputs[i][j];
val = LLVMBuildLoad(gallivm->builder, addr, "");
val = ac_build_clamp(&ctx->ac, val);
LLVMBuildStore(gallivm->builder, val, addr);
@@ -3257,21 +3249,21 @@ static void clock_emit(
emit_data->output[1] =
LLVMBuildExtractElement(gallivm->builder, tmp, ctx->i32_1, "");
}
static LLVMValueRef
shader_buffer_fetch_rsrc(struct si_shader_context *ctx,
const struct tgsi_full_src_register *reg)
{
LLVMValueRef index;
LLVMValueRef rsrc_ptr = LLVMGetParam(ctx->main_fn,
- SI_PARAM_SHADER_BUFFERS);
+ ctx->param_shader_buffers);
if (!reg->Register.Indirect)
index = LLVMConstInt(ctx->i32, reg->Register.Index, 0);
else
index = get_bounded_indirect_index(ctx, ®->Indirect,
reg->Register.Index,
SI_NUM_SHADER_BUFFERS);
return ac_build_indexed_load_const(&ctx->ac, rsrc_ptr, index);
}
@@ -3354,21 +3346,21 @@ static LLVMValueRef load_image_desc(struct
si_shader_context *ctx,
*/
static void
image_fetch_rsrc(
struct lp_build_tgsi_context *bld_base,
const struct tgsi_full_src_register *image,
bool is_store, unsigned target,
LLVMValueRef *rsrc)
{
struct si_shader_context *ctx = si_shader_context(bld_base);
LLVMValueRef rsrc_ptr = LLVMGetParam(ctx->main_fn,
- SI_PARAM_IMAGES);
+ ctx->param_images);
LLVMValueRef index;
bool dcc_off = is_store;
assert(image->Register.File == TGSI_FILE_IMAGE);
if (!image->Register.Indirect) {
const struct tgsi_shader_info *info = bld_base->info;
unsigned images_writemask = info->images_store |
info->images_atomic;
@@ -4366,21 +4358,21 @@ static LLVMValueRef sici_fix_sampler_aniso(struct
si_shader_context *ctx,
return LLVMBuildInsertElement(builder, samp, samp0,
ctx->i32_0, "");
}
static void tex_fetch_ptrs(
struct lp_build_tgsi_context *bld_base,
struct lp_build_emit_data *emit_data,
LLVMValueRef *res_ptr, LLVMValueRef *samp_ptr, LLVMValueRef *fmask_ptr)
{
struct si_shader_context *ctx = si_shader_context(bld_base);
- LLVMValueRef list = LLVMGetParam(ctx->main_fn, SI_PARAM_SAMPLERS);
+ LLVMValueRef list = LLVMGetParam(ctx->main_fn, ctx->param_samplers);
const struct tgsi_full_instruction *inst = emit_data->inst;
const struct tgsi_full_src_register *reg;
unsigned target = inst->Texture.Texture;
unsigned sampler_src;
LLVMValueRef index;
sampler_src = emit_data->inst->Instruction.NumSrcRegs - 1;
reg = &emit_data->inst->Src[sampler_src];
if (reg->Register.Indirect) {
@@ -5371,21 +5363,21 @@ static void si_llvm_emit_vertex(
struct lp_build_tgsi_context *bld_base,
struct lp_build_emit_data *emit_data)
{
struct si_shader_context *ctx = si_shader_context(bld_base);
struct lp_build_context *uint = &bld_base->uint_bld;
struct si_shader *shader = ctx->shader;
struct tgsi_shader_info *info = &shader->selector->info;
struct gallivm_state *gallivm = &ctx->gallivm;
struct lp_build_if_state if_state;
LLVMValueRef soffset = LLVMGetParam(ctx->main_fn,
- SI_PARAM_GS2VS_OFFSET);
+ ctx->param_gs2vs_offset);
LLVMValueRef gs_next_vertex;
LLVMValueRef can_emit, kill;
unsigned chan, offset;
int i;
unsigned stream;
stream = si_llvm_get_stream(bld_base, emit_data);
/* Write vertex attribute values to GSVS ring */
gs_next_vertex = LLVMBuildLoad(gallivm->builder,
@@ -5443,38 +5435,38 @@ static void si_llvm_emit_vertex(
}
}
gs_next_vertex = lp_build_add(uint, gs_next_vertex,
ctx->i32_1);
LLVMBuildStore(gallivm->builder, gs_next_vertex,
ctx->gs_next_vertex[stream]);
/* Signal vertex emission */
ac_build_sendmsg(&ctx->ac, AC_SENDMSG_GS_OP_EMIT | AC_SENDMSG_GS |
(stream << 8),
- LLVMGetParam(ctx->main_fn, SI_PARAM_GS_WAVE_ID));
+ LLVMGetParam(ctx->main_fn, ctx->param_gs_wave_id));
if (!use_kill)
lp_build_endif(&if_state);
}
/* Cut one primitive from the geometry shader */
static void si_llvm_emit_primitive(
const struct lp_build_tgsi_action *action,
struct lp_build_tgsi_context *bld_base,
struct lp_build_emit_data *emit_data)
{
struct si_shader_context *ctx = si_shader_context(bld_base);
unsigned stream;
/* Signal primitive cut */
stream = si_llvm_get_stream(bld_base, emit_data);
ac_build_sendmsg(&ctx->ac, AC_SENDMSG_GS_OP_CUT | AC_SENDMSG_GS |
(stream << 8),
- LLVMGetParam(ctx->main_fn, SI_PARAM_GS_WAVE_ID));
+ LLVMGetParam(ctx->main_fn, ctx->param_gs_wave_id));
}
static void si_llvm_emit_barrier(const struct lp_build_tgsi_action *action,
struct lp_build_tgsi_context *bld_base,
struct lp_build_emit_data *emit_data)
{
struct si_shader_context *ctx = si_shader_context(bld_base);
struct gallivm_state *gallivm = &ctx->gallivm;
/* SI only (thanks to a hw bug workaround):
@@ -5631,49 +5623,53 @@ static unsigned si_get_max_workgroup_size(struct
si_shader *shader)
return max_work_group_size;
}
static void create_function(struct si_shader_context *ctx)
{
struct lp_build_tgsi_context *bld_base = &ctx->bld_base;
struct gallivm_state *gallivm = &ctx->gallivm;
struct si_shader *shader = ctx->shader;
LLVMTypeRef params[SI_NUM_PARAMS + SI_MAX_ATTRIBS], v3i32;
LLVMTypeRef returns[16+32*4];
- unsigned i, last_sgpr, num_params, num_return_sgprs;
+ unsigned i, last_sgpr, num_params = 0, num_return_sgprs;
unsigned num_returns = 0;
unsigned num_prolog_vgprs = 0;
v3i32 = LLVMVectorType(ctx->i32, 3);
- params[SI_PARAM_RW_BUFFERS] = const_array(ctx->v16i8,
SI_NUM_RW_BUFFERS);
- params[SI_PARAM_CONST_BUFFERS] = const_array(ctx->v16i8,
SI_NUM_CONST_BUFFERS);
- params[SI_PARAM_SAMPLERS] = const_array(ctx->v8i32, SI_NUM_SAMPLERS);
- params[SI_PARAM_IMAGES] = const_array(ctx->v8i32, SI_NUM_IMAGES);
- params[SI_PARAM_SHADER_BUFFERS] = const_array(ctx->v4i32,
SI_NUM_SHADER_BUFFERS);
+ params[ctx->param_rw_buffers = num_params++] =
+ const_array(ctx->v16i8, SI_NUM_RW_BUFFERS);
+ params[ctx->param_const_buffers = num_params++] =
+ const_array(ctx->v16i8, SI_NUM_CONST_BUFFERS);
+ params[ctx->param_samplers = num_params++] =
+ const_array(ctx->v8i32, SI_NUM_SAMPLERS);
+ params[ctx->param_images = num_params++] =
+ const_array(ctx->v8i32, SI_NUM_IMAGES);
+ params[ctx->param_shader_buffers = num_params++] =
+ const_array(ctx->v4i32, SI_NUM_SHADER_BUFFERS);
switch (ctx->type) {
case PIPE_SHADER_VERTEX:
- params[SI_PARAM_VERTEX_BUFFERS] = const_array(ctx->v16i8,
SI_MAX_ATTRIBS);
- params[SI_PARAM_BASE_VERTEX] = ctx->i32;
- params[SI_PARAM_START_INSTANCE] = ctx->i32;
- params[SI_PARAM_DRAWID] = ctx->i32;
- params[SI_PARAM_VS_STATE_BITS] = ctx->i32;
- num_params = SI_PARAM_VS_STATE_BITS+1;
+ params[ctx->param_vertex_buffers = num_params++] =
+ const_array(ctx->v16i8, SI_NUM_VERTEX_BUFFERS);
+ params[ctx->param_base_vertex = num_params++] = ctx->i32;
+ params[ctx->param_start_instance = num_params++] = ctx->i32;
+ params[ctx->param_draw_id = num_params++] = ctx->i32;
+ params[ctx->param_vs_state_bits = num_params++] = ctx->i32;
if (shader->key.as_es) {
params[ctx->param_es2gs_offset = num_params++] =
ctx->i32;
} else if (shader->key.as_ls) {
/* no extra parameters */
} else {
- if (shader->is_gs_copy_shader) {
- num_params = SI_PARAM_RW_BUFFERS+1;
- }
+ if (shader->is_gs_copy_shader)
+ num_params = ctx->param_rw_buffers + 1;
/* The locations of the other parameters are assigned
dynamically. */
declare_streamout_params(ctx, &shader->selector->so,
params, ctx->i32, &num_params);
}
last_sgpr = num_params-1;
/* VGPRs */
params[ctx->param_vertex_id = num_params++] = ctx->i32;
@@ -5691,86 +5687,83 @@ static void create_function(struct si_shader_context
*ctx)
num_prolog_vgprs += shader->selector->info.num_inputs;
/* PrimitiveID output. */
if (!shader->key.as_es && !shader->key.as_ls)
for (i = 0; i <= VS_EPILOG_PRIMID_LOC; i++)
returns[num_returns++] = ctx->f32;
}
break;
case PIPE_SHADER_TESS_CTRL:
- params[SI_PARAM_TCS_OFFCHIP_LAYOUT] = ctx->i32;
- params[SI_PARAM_TCS_OUT_OFFSETS] = ctx->i32;
- params[SI_PARAM_TCS_OUT_LAYOUT] = ctx->i32;
- params[SI_PARAM_TCS_IN_LAYOUT] = ctx->i32;
- params[ctx->param_oc_lds = SI_PARAM_TCS_OC_LDS] = ctx->i32;
- params[SI_PARAM_TESS_FACTOR_OFFSET] = ctx->i32;
- last_sgpr = SI_PARAM_TESS_FACTOR_OFFSET;
+ params[ctx->param_tcs_offchip_layout = num_params++] = ctx->i32;
+ params[ctx->param_tcs_out_lds_offsets = num_params++] =
ctx->i32;
+ params[ctx->param_tcs_out_lds_layout = num_params++] = ctx->i32;
+ params[ctx->param_vs_state_bits = num_params++] = ctx->i32;
+ params[ctx->param_tcs_offchip_offset = num_params++] = ctx->i32;
+ params[ctx->param_tcs_factor_offset = num_params++] = ctx->i32;
+ last_sgpr = num_params - 1;
/* VGPRs */
- params[SI_PARAM_PATCH_ID] = ctx->i32;
- params[SI_PARAM_REL_IDS] = ctx->i32;
- num_params = SI_PARAM_REL_IDS+1;
+ params[ctx->param_tcs_patch_id = num_params++] = ctx->i32;
+ params[ctx->param_tcs_rel_ids = num_params++] = ctx->i32;
- /* SI_PARAM_TCS_OC_LDS and PARAM_TESS_FACTOR_OFFSET are
+ /* 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 < 3; i++)
returns[num_returns++] = ctx->f32; /* VGPRs */
break;
case PIPE_SHADER_TESS_EVAL:
- params[SI_PARAM_TCS_OFFCHIP_LAYOUT] = ctx->i32;
- num_params = SI_PARAM_TCS_OFFCHIP_LAYOUT+1;
+ params[ctx->param_tcs_offchip_layout = num_params++] = ctx->i32;
if (shader->key.as_es) {
- params[ctx->param_oc_lds = num_params++] = ctx->i32;
+ params[ctx->param_tcs_offchip_offset = num_params++] =
ctx->i32;
params[num_params++] = ctx->i32;
params[ctx->param_es2gs_offset = num_params++] =
ctx->i32;
} else {
params[num_params++] = ctx->i32;
declare_streamout_params(ctx, &shader->selector->so,
params, ctx->i32, &num_params);
- params[ctx->param_oc_lds = num_params++] = ctx->i32;
+ params[ctx->param_tcs_offchip_offset = num_params++] =
ctx->i32;
}
last_sgpr = num_params - 1;
/* VGPRs */
params[ctx->param_tes_u = num_params++] = ctx->f32;
params[ctx->param_tes_v = num_params++] = ctx->f32;
params[ctx->param_tes_rel_patch_id = num_params++] = ctx->i32;
params[ctx->param_tes_patch_id = num_params++] = ctx->i32;
/* PrimitiveID output. */
if (!shader->key.as_es)
for (i = 0; i <= VS_EPILOG_PRIMID_LOC; i++)
returns[num_returns++] = ctx->f32;
break;
case PIPE_SHADER_GEOMETRY:
- params[SI_PARAM_GS2VS_OFFSET] = ctx->i32;
- params[SI_PARAM_GS_WAVE_ID] = ctx->i32;
- last_sgpr = SI_PARAM_GS_WAVE_ID;
+ params[ctx->param_gs2vs_offset = num_params++] = ctx->i32;
+ params[ctx->param_gs_wave_id = num_params++] = ctx->i32;
+ last_sgpr = num_params - 1;
/* VGPRs */
- params[SI_PARAM_VTX0_OFFSET] = ctx->i32;
- params[SI_PARAM_VTX1_OFFSET] = ctx->i32;
- params[SI_PARAM_PRIMITIVE_ID] = ctx->i32;
- params[SI_PARAM_VTX2_OFFSET] = ctx->i32;
- params[SI_PARAM_VTX3_OFFSET] = ctx->i32;
- params[SI_PARAM_VTX4_OFFSET] = ctx->i32;
- params[SI_PARAM_VTX5_OFFSET] = ctx->i32;
- params[SI_PARAM_GS_INSTANCE_ID] = ctx->i32;
- num_params = SI_PARAM_GS_INSTANCE_ID+1;
+ params[ctx->param_gs_vtx0_offset = num_params++] = ctx->i32;
+ params[ctx->param_gs_vtx1_offset = num_params++] = ctx->i32;
+ params[ctx->param_gs_prim_id = num_params++] = ctx->i32;
+ params[ctx->param_gs_vtx2_offset = num_params++] = ctx->i32;
+ params[ctx->param_gs_vtx3_offset = num_params++] = ctx->i32;
+ params[ctx->param_gs_vtx4_offset = num_params++] = ctx->i32;
+ params[ctx->param_gs_vtx5_offset = num_params++] = ctx->i32;
+ params[ctx->param_gs_instance_id = num_params++] = ctx->i32;
break;
case PIPE_SHADER_FRAGMENT:
params[SI_PARAM_ALPHA_REF] = ctx->f32;
params[SI_PARAM_PRIM_MASK] = ctx->i32;
last_sgpr = SI_PARAM_PRIM_MASK;
params[SI_PARAM_PERSP_SAMPLE] = ctx->v2i32;
params[SI_PARAM_PERSP_CENTER] = ctx->v2i32;
params[SI_PARAM_PERSP_CENTROID] = ctx->v2i32;
params[SI_PARAM_PERSP_PULL_MODEL] = v3i32;
@@ -5893,21 +5886,21 @@ static void create_function(struct si_shader_context
*ctx)
/**
* Load ESGS and GSVS ring buffer resource descriptors and save the variables
* for later use.
*/
static void preload_ring_buffers(struct si_shader_context *ctx)
{
struct gallivm_state *gallivm = &ctx->gallivm;
LLVMBuilderRef builder = gallivm->builder;
LLVMValueRef buf_ptr = LLVMGetParam(ctx->main_fn,
- SI_PARAM_RW_BUFFERS);
+ ctx->param_rw_buffers);
if ((ctx->type == PIPE_SHADER_VERTEX &&
ctx->shader->key.as_es) ||
(ctx->type == PIPE_SHADER_TESS_EVAL &&
ctx->shader->key.as_es) ||
ctx->type == PIPE_SHADER_GEOMETRY) {
unsigned ring =
ctx->type == PIPE_SHADER_GEOMETRY ? SI_GS_RING_ESGS
: SI_ES_RING_ESGS;
LLVMValueRef offset = LLVMConstInt(ctx->i32, ring, 0);
@@ -8098,36 +8091,36 @@ static bool si_shader_select_tes_parts(struct si_screen
*sscreen,
* Compile the TCS epilog function. This writes tesselation factors to memory
* based on the output primitive type of the tesselator (determined by TES).
*/
static void si_build_tcs_epilog_function(struct si_shader_context *ctx,
union si_shader_part_key *key)
{
struct gallivm_state *gallivm = &ctx->gallivm;
struct lp_build_tgsi_context *bld_base = &ctx->bld_base;
LLVMTypeRef params[16];
LLVMValueRef func;
- int last_sgpr, num_params;
+ int last_sgpr, num_params = 0;
/* Declare inputs. Only RW_BUFFERS and TESS_FACTOR_OFFSET are used. */
- params[SI_PARAM_RW_BUFFERS] = const_array(ctx->v16i8,
SI_NUM_RW_BUFFERS);
- params[SI_PARAM_CONST_BUFFERS] = ctx->i64;
- params[SI_PARAM_SAMPLERS] = ctx->i64;
- params[SI_PARAM_IMAGES] = ctx->i64;
- params[SI_PARAM_SHADER_BUFFERS] = ctx->i64;
- params[SI_PARAM_TCS_OFFCHIP_LAYOUT] = ctx->i32;
- params[SI_PARAM_TCS_OUT_OFFSETS] = ctx->i32;
- params[SI_PARAM_TCS_OUT_LAYOUT] = ctx->i32;
- params[SI_PARAM_TCS_IN_LAYOUT] = ctx->i32;
- params[ctx->param_oc_lds = SI_PARAM_TCS_OC_LDS] = ctx->i32;
- params[SI_PARAM_TESS_FACTOR_OFFSET] = ctx->i32;
- last_sgpr = SI_PARAM_TESS_FACTOR_OFFSET;
- num_params = last_sgpr + 1;
+ params[ctx->param_rw_buffers = num_params++] =
+ const_array(ctx->v16i8, SI_NUM_RW_BUFFERS);
+ params[ctx->param_const_buffers = num_params++] = ctx->i64;
+ params[ctx->param_samplers = num_params++] = ctx->i64;
+ params[ctx->param_images = num_params++] = ctx->i64;
+ params[ctx->param_shader_buffers = num_params++] = ctx->i64;
+ params[ctx->param_tcs_offchip_layout = num_params++] = ctx->i32;
+ params[ctx->param_tcs_out_lds_offsets = num_params++] = ctx->i32;
+ params[ctx->param_tcs_out_lds_layout = num_params++] = ctx->i32;
+ params[ctx->param_vs_state_bits = num_params++] = ctx->i32;
+ params[ctx->param_tcs_offchip_offset = num_params++] = ctx->i32;
+ params[ctx->param_tcs_factor_offset = num_params++] = ctx->i32;
+ last_sgpr = num_params - 1;
params[num_params++] = ctx->i32; /* patch index within the wave
(REL_PATCH_ID) */
params[num_params++] = ctx->i32; /* invocation ID within the patch */
params[num_params++] = ctx->i32; /* LDS offset where tess factors
should be loaded from */
/* Create the function. */
si_create_function(ctx, "tcs_epilog", NULL, 0, params, num_params,
last_sgpr);
declare_tess_lds(ctx);
func = ctx->main_fn;
@@ -8443,29 +8436,30 @@ static void si_build_ps_prolog_function(struct
si_shader_context *ctx,
* Build the pixel shader epilog function. This handles everything that must be
* emulated for pixel shader exports. (alpha-test, format conversions, etc)
*/
static void si_build_ps_epilog_function(struct si_shader_context *ctx,
union si_shader_part_key *key)
{
struct gallivm_state *gallivm = &ctx->gallivm;
struct lp_build_tgsi_context *bld_base = &ctx->bld_base;
LLVMTypeRef params[16+8*4+3];
LLVMValueRef depth = NULL, stencil = NULL, samplemask = NULL;
- int last_sgpr, num_params, i;
+ int last_sgpr, num_params = 0, i;
struct si_ps_exports exp = {};
/* Declare input SGPRs. */
- params[SI_PARAM_RW_BUFFERS] = ctx->i64;
- params[SI_PARAM_CONST_BUFFERS] = ctx->i64;
- params[SI_PARAM_SAMPLERS] = ctx->i64;
- params[SI_PARAM_IMAGES] = ctx->i64;
- params[SI_PARAM_SHADER_BUFFERS] = ctx->i64;
+ params[ctx->param_rw_buffers = num_params++] = ctx->i64;
+ params[ctx->param_const_buffers = num_params++] = ctx->i64;
+ params[ctx->param_samplers = num_params++] = ctx->i64;
+ params[ctx->param_images = num_params++] = ctx->i64;
+ params[ctx->param_shader_buffers = num_params++] = ctx->i64;
+ assert(num_params == SI_PARAM_ALPHA_REF);
params[SI_PARAM_ALPHA_REF] = ctx->f32;
last_sgpr = SI_PARAM_ALPHA_REF;
/* Declare input VGPRs. */
num_params = (last_sgpr + 1) +
util_bitcount(key->ps_epilog.colors_written) * 4 +
key->ps_epilog.writes_z +
key->ps_epilog.writes_stencil +
key->ps_epilog.writes_samplemask;
diff --git a/src/gallium/drivers/radeonsi/si_shader.h
b/src/gallium/drivers/radeonsi/si_shader.h
index fa6f9af..1fee044 100644
--- a/src/gallium/drivers/radeonsi/si_shader.h
+++ b/src/gallium/drivers/radeonsi/si_shader.h
@@ -136,79 +136,21 @@ enum {
SI_PS_NUM_USER_SGPR,
/* CS only */
SI_SGPR_GRID_SIZE = SI_NUM_RESOURCE_SGPRS,
SI_SGPR_BLOCK_SIZE = SI_SGPR_GRID_SIZE + 3,
SI_CS_NUM_USER_SGPR = SI_SGPR_BLOCK_SIZE + 3
};
/* LLVM function parameter indices */
enum {
- SI_PARAM_RW_BUFFERS,
- SI_PARAM_CONST_BUFFERS,
- SI_PARAM_SAMPLERS,
- SI_PARAM_IMAGES,
- SI_PARAM_SHADER_BUFFERS,
- SI_NUM_RESOURCE_PARAMS,
-
- /* VS only parameters */
- SI_PARAM_VERTEX_BUFFERS = SI_NUM_RESOURCE_PARAMS,
- SI_PARAM_BASE_VERTEX,
- SI_PARAM_START_INSTANCE,
- SI_PARAM_DRAWID,
- SI_PARAM_VS_STATE_BITS,
-
- /* Layout of TCS outputs in the offchip buffer
- * [0:8] = the number of patches per threadgroup.
- * [9:15] = the number of output vertices per patch.
- * [16:31] = the offset of per patch attributes in the buffer in
bytes.
- */
- SI_PARAM_TCS_OFFCHIP_LAYOUT = SI_NUM_RESOURCE_PARAMS, /* for TCS & TES
*/
-
- /* TCS only parameters. */
-
- /* Offsets where TCS outputs and TCS patch outputs live in LDS:
- * [0:15] = TCS output patch0 offset / 16, max = NUM_PATCHES * 32 * 32
- * [16:31] = TCS output patch0 offset for per-patch / 16, max =
NUM_PATCHES*32*32* + 32*32
- */
- SI_PARAM_TCS_OUT_OFFSETS,
-
- /* Layout of TCS outputs / TES inputs:
- * [0:12] = stride between output patches in dwords, num_outputs *
num_vertices * 4, max = 32*32*4
- * [13:20] = stride between output vertices in dwords = num_inputs *
4, max = 32*4
- * [26:31] = gl_PatchVerticesIn, max = 32
- */
- SI_PARAM_TCS_OUT_LAYOUT,
-
- /* Layout of LS outputs / TCS inputs
- * [8:20] = stride between patches in dwords = num_inputs *
num_vertices * 4, max = 32*32*4
- * [24:31] = stride between vertices in dwords = num_inputs * 4, max
= 32*4
- * (same layout as SI_PARAM_VS_STATE_BITS)
- */
- SI_PARAM_TCS_IN_LAYOUT,
-
- SI_PARAM_TCS_OC_LDS,
- SI_PARAM_TESS_FACTOR_OFFSET,
- SI_PARAM_PATCH_ID,
- SI_PARAM_REL_IDS,
-
- /* GS only parameters */
- SI_PARAM_GS2VS_OFFSET = SI_NUM_RESOURCE_PARAMS,
- SI_PARAM_GS_WAVE_ID,
- SI_PARAM_VTX0_OFFSET,
- SI_PARAM_VTX1_OFFSET,
- SI_PARAM_PRIMITIVE_ID,
- SI_PARAM_VTX2_OFFSET,
- SI_PARAM_VTX3_OFFSET,
- SI_PARAM_VTX4_OFFSET,
- SI_PARAM_VTX5_OFFSET,
- SI_PARAM_GS_INSTANCE_ID,
+ SI_NUM_RESOURCE_PARAMS = 5,
/* PS only parameters */
SI_PARAM_ALPHA_REF = SI_NUM_RESOURCE_PARAMS,
SI_PARAM_PRIM_MASK,
SI_PARAM_PERSP_SAMPLE,
SI_PARAM_PERSP_CENTER,
SI_PARAM_PERSP_CENTROID,
SI_PARAM_PERSP_PULL_MODEL,
SI_PARAM_LINEAR_SAMPLE,
SI_PARAM_LINEAR_CENTER,
diff --git a/src/gallium/drivers/radeonsi/si_shader_internal.h
b/src/gallium/drivers/radeonsi/si_shader_internal.h
index 3f856c4..812472f 100644
--- a/src/gallium/drivers/radeonsi/si_shader_internal.h
+++ b/src/gallium/drivers/radeonsi/si_shader_internal.h
@@ -100,34 +100,95 @@ struct si_shader_context {
unsigned flow_depth_max;
struct tgsi_array_info *temp_arrays;
LLVMValueRef *temp_array_allocas;
LLVMValueRef undef_alloca;
LLVMValueRef main_fn;
LLVMTypeRef return_type;
- int param_streamout_config;
- int param_streamout_write_index;
- int param_streamout_offset[4];
+ /* Parameter indices for LLVMGetParam. */
+ int param_rw_buffers;
+ int param_const_buffers;
+ int param_samplers;
+ int param_images;
+ int param_shader_buffers;
+ /* API VS */
+ int param_vertex_buffers;
+ int param_base_vertex;
+ int param_start_instance;
+ int param_draw_id;
int param_vertex_id;
int param_rel_auto_id;
int param_vs_prim_id;
int param_instance_id;
int param_vertex_index0;
+ /* VS states and layout of LS outputs / TCS inputs at the end
+ * [0] = clamp vertex color
+ * [1] = indexed
+ * [8:20] = stride between patches in DW = num_inputs * num_vertices
* 4
+ * max = 32*32*4
+ * [24:31] = stride between vertices in DW = num_inputs * 4
+ * max = 32*4
+ */
+ int param_vs_state_bits;
+ /* HW VS */
+ int param_streamout_config;
+ int param_streamout_write_index;
+ int param_streamout_offset[4];
+
+ /* API TCS & TES */
+ /* Layout of TCS outputs in the offchip buffer
+ * [0:8] = the number of patches per threadgroup.
+ * [9:15] = the number of output vertices per patch.
+ * [16:31] = the offset of per patch attributes in the buffer in
bytes. */
+ int param_tcs_offchip_layout;
+
+ /* API TCS */
+ /* Offsets where TCS outputs and TCS patch outputs live in LDS:
+ * [0:15] = TCS output patch0 offset / 16, max = NUM_PATCHES * 32 * 32
+ * [16:31] = TCS output patch0 offset for per-patch / 16
+ * max = NUM_PATCHES*32*32* + 32*32
+ */
+ int param_tcs_out_lds_offsets;
+ /* Layout of TCS outputs / TES inputs:
+ * [0:12] = stride between output patches in DW, num_outputs *
num_vertices * 4
+ * max = 32*32*4
+ * [13:20] = stride between output vertices in DW = num_inputs * 4
+ * max = 32*4
+ * [26:31] = gl_PatchVerticesIn, max = 32
+ */
+ int param_tcs_out_lds_layout;
+ int param_tcs_offchip_offset;
+ int param_tcs_factor_offset;
+ int param_tcs_patch_id;
+ int param_tcs_rel_ids;
+
+ /* API TES */
int param_tes_u;
int param_tes_v;
int param_tes_rel_patch_id;
int param_tes_patch_id;
+ /* HW ES */
int param_es2gs_offset;
- int param_oc_lds;
+ /* API GS */
+ int param_gs2vs_offset;
+ int param_gs_wave_id;
+ int param_gs_vtx0_offset;
+ int param_gs_vtx1_offset;
+ int param_gs_prim_id;
+ int param_gs_vtx2_offset;
+ int param_gs_vtx3_offset;
+ int param_gs_vtx4_offset;
+ int param_gs_vtx5_offset;
+ int param_gs_instance_id;
LLVMTargetMachineRef tm;
unsigned range_md_kind;
unsigned fpmath_md_kind;
LLVMValueRef fpmath_md_2p5_ulp;
/* Preloaded descriptors. */
LLVMValueRef esgs_ring;
LLVMValueRef gsvs_ring[4];
--
2.7.4
_______________________________________________
mesa-dev mailing list
mesa-dev@lists.freedesktop.org
https://lists.freedesktop.org/mailman/listinfo/mesa-dev
