[Mesa-dev] [PATCH 3/4] radeonsi: make const_array externally accessible
Nicolai Hähnle
nhaehnle at gmail.com
Mon May 15 21:43:59 UTC 2017
From: Nicolai Hähnle <nicolai.haehnle at amd.com>
---
src/gallium/drivers/radeonsi/si_shader.c | 26 +++++++++++------------
src/gallium/drivers/radeonsi/si_shader_internal.h | 2 ++
2 files changed, 15 insertions(+), 13 deletions(-)
diff --git a/src/gallium/drivers/radeonsi/si_shader.c b/src/gallium/drivers/radeonsi/si_shader.c
index 1001b27..56a86cd 100644
--- a/src/gallium/drivers/radeonsi/si_shader.c
+++ b/src/gallium/drivers/radeonsi/si_shader.c
@@ -3522,39 +3522,39 @@ static LLVMValueRef force_dcc_off(struct si_shader_context *ctx,
LLVMValueRef i32_6 = LLVMConstInt(ctx->i32, 6, 0);
LLVMValueRef i32_C = LLVMConstInt(ctx->i32, C_008F28_COMPRESSION_EN, 0);
LLVMValueRef tmp;
tmp = LLVMBuildExtractElement(builder, rsrc, i32_6, "");
tmp = LLVMBuildAnd(builder, tmp, i32_C, "");
return LLVMBuildInsertElement(builder, rsrc, tmp, i32_6, "");
}
}
-static LLVMTypeRef const_array(LLVMTypeRef elem_type, int num_elements)
+LLVMTypeRef si_const_array(LLVMTypeRef elem_type, int num_elements)
{
return LLVMPointerType(LLVMArrayType(elem_type, num_elements),
CONST_ADDR_SPACE);
}
static LLVMValueRef load_image_desc(struct si_shader_context *ctx,
LLVMValueRef list, LLVMValueRef index,
unsigned target)
{
LLVMBuilderRef builder = ctx->gallivm.builder;
if (target == TGSI_TEXTURE_BUFFER) {
index = LLVMBuildMul(builder, index,
LLVMConstInt(ctx->i32, 2, 0), "");
index = LLVMBuildAdd(builder, index,
ctx->i32_1, "");
list = LLVMBuildPointerCast(builder, list,
- const_array(ctx->v4i32, 0), "");
+ si_const_array(ctx->v4i32, 0), "");
}
return ac_build_indexed_load_const(&ctx->ac, list, index);
}
/**
* Load the resource descriptor for \p image.
*/
static void
image_fetch_rsrc(
@@ -4514,33 +4514,33 @@ static LLVMValueRef load_sampler_desc(struct si_shader_context *ctx,
switch (type) {
case DESC_IMAGE:
/* The image is at [0:7]. */
index = LLVMBuildMul(builder, index, LLVMConstInt(ctx->i32, 2, 0), "");
break;
case DESC_BUFFER:
/* The buffer is in [4:7]. */
index = LLVMBuildMul(builder, index, LLVMConstInt(ctx->i32, 4, 0), "");
index = LLVMBuildAdd(builder, index, ctx->i32_1, "");
list = LLVMBuildPointerCast(builder, list,
- const_array(ctx->v4i32, 0), "");
+ si_const_array(ctx->v4i32, 0), "");
break;
case DESC_FMASK:
/* The FMASK is at [8:15]. */
index = LLVMBuildMul(builder, index, LLVMConstInt(ctx->i32, 2, 0), "");
index = LLVMBuildAdd(builder, index, ctx->i32_1, "");
break;
case DESC_SAMPLER:
/* The sampler state is at [12:15]. */
index = LLVMBuildMul(builder, index, LLVMConstInt(ctx->i32, 4, 0), "");
index = LLVMBuildAdd(builder, index, LLVMConstInt(ctx->i32, 3, 0), "");
list = LLVMBuildPointerCast(builder, list,
- const_array(ctx->v4i32, 0), "");
+ si_const_array(ctx->v4i32, 0), "");
break;
}
return ac_build_indexed_load_const(&ctx->ac, list, index);
}
/* Disable anisotropic filtering if BASE_LEVEL == LAST_LEVEL.
*
* SI-CI:
* If BASE_LEVEL == LAST_LEVEL, the shader must disable anisotropic
@@ -5847,48 +5847,48 @@ static unsigned si_get_max_workgroup_size(const struct si_shader *shader)
max_work_group_size = SI_MAX_VARIABLE_THREADS_PER_BLOCK;
}
return max_work_group_size;
}
static void declare_per_stage_desc_pointers(struct si_shader_context *ctx,
LLVMTypeRef *params,
unsigned *num_params,
bool assign_params)
{
- params[(*num_params)++] = const_array(ctx->v4i32, SI_NUM_CONST_BUFFERS);
- params[(*num_params)++] = const_array(ctx->v8i32, SI_NUM_SAMPLERS);
- params[(*num_params)++] = const_array(ctx->v8i32, SI_NUM_IMAGES);
- params[(*num_params)++] = const_array(ctx->v4i32, SI_NUM_SHADER_BUFFERS);
+ params[(*num_params)++] = si_const_array(ctx->v4i32, SI_NUM_CONST_BUFFERS);
+ params[(*num_params)++] = si_const_array(ctx->v8i32, SI_NUM_SAMPLERS);
+ params[(*num_params)++] = si_const_array(ctx->v8i32, SI_NUM_IMAGES);
+ params[(*num_params)++] = si_const_array(ctx->v4i32, SI_NUM_SHADER_BUFFERS);
if (assign_params) {
ctx->param_const_buffers = *num_params - 4;
ctx->param_samplers = *num_params - 3;
ctx->param_images = *num_params - 2;
ctx->param_shader_buffers = *num_params - 1;
}
}
static void declare_default_desc_pointers(struct si_shader_context *ctx,
LLVMTypeRef *params,
unsigned *num_params)
{
params[ctx->param_rw_buffers = (*num_params)++] =
- const_array(ctx->v4i32, SI_NUM_RW_BUFFERS);
+ si_const_array(ctx->v4i32, SI_NUM_RW_BUFFERS);
declare_per_stage_desc_pointers(ctx, params, num_params, true);
}
static void declare_vs_specific_input_sgprs(struct si_shader_context *ctx,
LLVMTypeRef *params,
unsigned *num_params)
{
params[ctx->param_vertex_buffers = (*num_params)++] =
- const_array(ctx->v4i32, SI_NUM_VERTEX_BUFFERS);
+ si_const_array(ctx->v4i32, 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;
}
static void declare_vs_input_vgprs(struct si_shader_context *ctx,
LLVMTypeRef *params, unsigned *num_params,
unsigned *num_prolog_vgprs)
{
@@ -5996,21 +5996,21 @@ static void create_function(struct si_shader_context *ctx)
*/
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 SI_SHADER_MERGED_VERTEX_TESSCTRL:
/* Merged stages have 8 system SGPRs at the beginning. */
params[ctx->param_rw_buffers = num_params++] = /* SPI_SHADER_USER_DATA_ADDR_LO_HS */
- const_array(ctx->v4i32, SI_NUM_RW_BUFFERS);
+ si_const_array(ctx->v4i32, SI_NUM_RW_BUFFERS);
params[ctx->param_tcs_offchip_offset = num_params++] = ctx->i32;
params[ctx->param_merged_wave_info = num_params++] = ctx->i32;
params[ctx->param_tcs_factor_offset = num_params++] = ctx->i32;
params[ctx->param_merged_scratch_offset = num_params++] = ctx->i32;
params[num_params++] = ctx->i32; /* unused */
params[num_params++] = ctx->i32; /* unused */
params[num_params++] = ctx->i32; /* unused */
params[num_params++] = ctx->i32; /* unused */
declare_per_stage_desc_pointers(ctx, params, &num_params,
@@ -6051,21 +6051,21 @@ static void create_function(struct si_shader_context *ctx)
for (i = 0; i <= 8 + GFX9_SGPR_TCS_FACTOR_ADDR_BASE64K; i++)
returns[num_returns++] = ctx->i32; /* SGPRs */
for (i = 0; i < 3; i++)
returns[num_returns++] = ctx->f32; /* VGPRs */
}
break;
case SI_SHADER_MERGED_VERTEX_OR_TESSEVAL_GEOMETRY:
/* Merged stages have 8 system SGPRs at the beginning. */
params[ctx->param_rw_buffers = num_params++] = /* SPI_SHADER_USER_DATA_ADDR_LO_GS */
- const_array(ctx->v4i32, SI_NUM_RW_BUFFERS);
+ si_const_array(ctx->v4i32, SI_NUM_RW_BUFFERS);
params[ctx->param_gs2vs_offset = num_params++] = ctx->i32;
params[ctx->param_merged_wave_info = num_params++] = ctx->i32;
params[ctx->param_tcs_offchip_offset = num_params++] = ctx->i32;
params[ctx->param_merged_scratch_offset = num_params++] = ctx->i32;
params[num_params++] = ctx->i32; /* unused (SPI_SHADER_PGM_LO/HI_GS << 8) */
params[num_params++] = ctx->i32; /* unused (SPI_SHADER_PGM_LO/HI_GS >> 24) */
params[num_params++] = ctx->i32; /* unused */
params[num_params++] = ctx->i32; /* unused */
declare_per_stage_desc_pointers(ctx, params, &num_params,
@@ -8683,21 +8683,21 @@ static void si_build_ps_prolog_function(struct si_shader_context *ctx,
unsigned pos = key->ps_prolog.num_input_sgprs +
key->ps_prolog.num_input_vgprs - 1;
LLVMValueRef ptr[2], list;
/* Get the pointer to rw buffers. */
ptr[0] = LLVMGetParam(func, SI_SGPR_RW_BUFFERS);
ptr[1] = LLVMGetParam(func, SI_SGPR_RW_BUFFERS_HI);
list = lp_build_gather_values(gallivm, ptr, 2);
list = LLVMBuildBitCast(gallivm->builder, list, ctx->i64, "");
list = LLVMBuildIntToPtr(gallivm->builder, list,
- const_array(ctx->v4i32, SI_NUM_RW_BUFFERS), "");
+ si_const_array(ctx->v4i32, SI_NUM_RW_BUFFERS), "");
si_llvm_emit_polygon_stipple(ctx, list, pos);
}
if (key->ps_prolog.states.bc_optimize_for_persp ||
key->ps_prolog.states.bc_optimize_for_linear) {
unsigned i, base = key->ps_prolog.num_input_sgprs;
LLVMValueRef center[2], centroid[2], tmp, bc_optimize;
/* The shader should do: if (PRIM_MASK[31]) CENTROID = CENTER;
diff --git a/src/gallium/drivers/radeonsi/si_shader_internal.h b/src/gallium/drivers/radeonsi/si_shader_internal.h
index 70004fa..82a672f 100644
--- a/src/gallium/drivers/radeonsi/si_shader_internal.h
+++ b/src/gallium/drivers/radeonsi/si_shader_internal.h
@@ -291,13 +291,15 @@ void si_llvm_emit_store(struct lp_build_tgsi_context *bld_base,
#define NOOP_WAITCNT 0xf7f
#define LGKM_CNT 0x07f
#define VM_CNT 0xf70
void si_emit_waitcnt(struct si_shader_context *ctx, unsigned simm16);
LLVMValueRef si_get_bounded_indirect_index(struct si_shader_context *ctx,
const struct tgsi_ind_register *ind,
int rel_index, unsigned num);
+LLVMTypeRef si_const_array(LLVMTypeRef elem_type, int num_elements);
+
void si_shader_context_init_alu(struct lp_build_tgsi_context *bld_base);
#endif
--
2.9.3
More information about the mesa-dev
mailing list