[Mesa-dev] [PATCH 3/3] radeonsi: access gallivm through ctx in most places
Nicolai Hähnle
nhaehnle at gmail.com
Tue Apr 4 14:48:43 UTC 2017
For the series:
Reviewed-by: Nicolai Hähnle <nicolai.haehnle at amd.com>
And FWIW, I like i32_0/1. It's shorter :)
On 03.04.2017 11:52, Marek Olšák wrote:
> From: Marek Olšák <marek.olsak at amd.com>
>
> ---
> src/gallium/drivers/radeonsi/si_shader.c | 116 ++++++++++-----------
> src/gallium/drivers/radeonsi/si_shader_tgsi_alu.c | 4 +-
> .../drivers/radeonsi/si_shader_tgsi_setup.c | 46 ++++----
> 3 files changed, 79 insertions(+), 87 deletions(-)
>
> diff --git a/src/gallium/drivers/radeonsi/si_shader.c b/src/gallium/drivers/radeonsi/si_shader.c
> index 0200172..29d3dd4 100644
> --- a/src/gallium/drivers/radeonsi/si_shader.c
> +++ b/src/gallium/drivers/radeonsi/si_shader.c
> @@ -296,21 +296,21 @@ get_tcs_out_current_patch_data_offset(struct si_shader_context *ctx)
> return LLVMBuildAdd(gallivm->builder, patch0_patch_data_offset,
> LLVMBuildMul(gallivm->builder, patch_stride,
> rel_patch_id, ""),
> "");
> }
>
> static LLVMValueRef get_instance_index_for_fetch(
> struct si_shader_context *ctx,
> unsigned param_start_instance, unsigned divisor)
> {
> - struct gallivm_state *gallivm = ctx->bld_base.base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
>
> LLVMValueRef result = LLVMGetParam(ctx->main_fn,
> ctx->param_instance_id);
>
> /* The division must be done before START_INSTANCE is added. */
> if (divisor > 1)
> result = LLVMBuildUDiv(gallivm->builder, result,
> LLVMConstInt(ctx->i32, divisor, 0), "");
>
> return LLVMBuildAdd(gallivm->builder, result,
> @@ -331,22 +331,21 @@ static LLVMValueRef extract_double_to_float(struct si_shader_context *ctx,
> LLVMValueRef value = LLVMBuildExtractElement(builder, dvec2, index, "");
> return LLVMBuildFPTrunc(builder, value, ctx->f32, "");
> }
>
> static void declare_input_vs(
> struct si_shader_context *ctx,
> unsigned input_index,
> const struct tgsi_full_declaration *decl,
> LLVMValueRef out[4])
> {
> - struct lp_build_context *base = &ctx->bld_base.base;
> - struct gallivm_state *gallivm = base->gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
>
> unsigned chan;
> unsigned fix_fetch;
> unsigned num_fetches;
> unsigned fetch_stride;
>
> LLVMValueRef t_list_ptr;
> LLVMValueRef t_offset;
> LLVMValueRef t_list;
> LLVMValueRef vertex_index;
> @@ -567,21 +566,21 @@ static LLVMValueRef get_primitive_id(struct lp_build_tgsi_context *bld_base,
> }
>
> /**
> * Return the value of tgsi_ind_register for indexing.
> * This is the indirect index with the constant offset added to it.
> */
> static LLVMValueRef get_indirect_index(struct si_shader_context *ctx,
> const struct tgsi_ind_register *ind,
> int rel_index)
> {
> - struct gallivm_state *gallivm = ctx->bld_base.base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> LLVMValueRef result;
>
> result = ctx->addrs[ind->Index][ind->Swizzle];
> result = LLVMBuildLoad(gallivm->builder, result, "");
> result = LLVMBuildAdd(gallivm->builder, result,
> LLVMConstInt(ctx->i32, rel_index, 0), "");
> return result;
> }
>
> /**
> @@ -607,21 +606,21 @@ static LLVMValueRef get_bounded_indirect_index(struct si_shader_context *ctx,
>
> /**
> * Calculate a dword address given an input or output register and a stride.
> */
> static LLVMValueRef get_dw_address(struct si_shader_context *ctx,
> const struct tgsi_full_dst_register *dst,
> const struct tgsi_full_src_register *src,
> LLVMValueRef vertex_dw_stride,
> LLVMValueRef base_addr)
> {
> - struct gallivm_state *gallivm = ctx->bld_base.base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> struct tgsi_shader_info *info = &ctx->shader->selector->info;
> ubyte *name, *index, *array_first;
> int first, param;
> struct tgsi_full_dst_register reg;
>
> /* Set the register description. The address computation is the same
> * for sources and destinations. */
> if (src) {
> reg.Register.File = src->Register.File;
> reg.Register.Index = src->Register.Index;
> @@ -706,21 +705,21 @@ static LLVMValueRef get_dw_address(struct si_shader_context *ctx,
> * - per patch attribute 0 of patch 1
> * ...
> *
> * Note that every attribute has 4 components.
> */
> 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->bld_base.base.gallivm;
> + 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);
> total_vertices = LLVMBuildMul(gallivm->builder, vertices_per_patch,
> num_patches, "");
>
> constant16 = LLVMConstInt(ctx->i32, 16, 0);
> if (vertex_index) {
> @@ -750,21 +749,21 @@ static LLVMValueRef get_tcs_tes_buffer_address(struct si_shader_context *ctx,
> 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,
> const struct tgsi_full_src_register *src)
> {
> - struct gallivm_state *gallivm = ctx->bld_base.base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> struct tgsi_shader_info *info = &ctx->shader->selector->info;
> ubyte *name, *index, *array_first;
> struct tgsi_full_src_register reg;
> LLVMValueRef vertex_index = NULL;
> LLVMValueRef param_index = NULL;
> unsigned param_index_base, param_base;
>
> reg = src ? *src : tgsi_full_src_register_from_dst(dst);
>
> if (reg.Register.Dimension) {
> @@ -814,21 +813,21 @@ static LLVMValueRef get_tcs_tes_buffer_address_from_reg(
> return get_tcs_tes_buffer_address(ctx, get_rel_patch_id(ctx),
> vertex_index, param_index);
> }
>
> static LLVMValueRef buffer_load(struct lp_build_tgsi_context *bld_base,
> enum tgsi_opcode_type type, unsigned swizzle,
> LLVMValueRef buffer, LLVMValueRef offset,
> LLVMValueRef base, bool readonly_memory)
> {
> struct si_shader_context *ctx = si_shader_context(bld_base);
> - struct gallivm_state *gallivm = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> LLVMValueRef value, value2;
> LLVMTypeRef llvm_type = tgsi2llvmtype(bld_base, type);
> LLVMTypeRef vec_type = LLVMVectorType(llvm_type, 4);
>
> if (swizzle == ~0) {
> value = ac_build_buffer_load(&ctx->ac, buffer, 4, NULL, base, offset,
> 0, 1, 0, readonly_memory);
>
> return LLVMBuildBitCast(gallivm->builder, value, vec_type, "");
> }
> @@ -856,30 +855,30 @@ static LLVMValueRef buffer_load(struct lp_build_tgsi_context *bld_base,
> *
> * \param type output value type
> * \param swizzle offset (typically 0..3); it can be ~0, which loads a vec4
> * \param dw_addr address in dwords
> */
> static LLVMValueRef lds_load(struct lp_build_tgsi_context *bld_base,
> enum tgsi_opcode_type type, unsigned swizzle,
> LLVMValueRef dw_addr)
> {
> struct si_shader_context *ctx = si_shader_context(bld_base);
> - struct gallivm_state *gallivm = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> LLVMValueRef value;
>
> if (swizzle == ~0) {
> LLVMValueRef values[TGSI_NUM_CHANNELS];
>
> for (unsigned chan = 0; chan < TGSI_NUM_CHANNELS; chan++)
> values[chan] = lds_load(bld_base, type, chan, dw_addr);
>
> - return lp_build_gather_values(bld_base->base.gallivm, values,
> + return lp_build_gather_values(gallivm, values,
> TGSI_NUM_CHANNELS);
> }
>
> dw_addr = lp_build_add(&bld_base->uint_bld, dw_addr,
> LLVMConstInt(ctx->i32, swizzle, 0));
>
> value = ac_build_indexed_load(&ctx->ac, ctx->lds, dw_addr, false);
> if (tgsi_type_is_64bit(type)) {
> LLVMValueRef value2;
> dw_addr = lp_build_add(&bld_base->uint_bld, dw_addr,
> @@ -897,21 +896,21 @@ static LLVMValueRef lds_load(struct lp_build_tgsi_context *bld_base,
> *
> * \param swizzle offset (typically 0..3)
> * \param dw_addr address in dwords
> * \param value value to store
> */
> static void lds_store(struct lp_build_tgsi_context *bld_base,
> unsigned swizzle, LLVMValueRef dw_addr,
> LLVMValueRef value)
> {
> struct si_shader_context *ctx = si_shader_context(bld_base);
> - struct gallivm_state *gallivm = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
>
> dw_addr = lp_build_add(&bld_base->uint_bld, dw_addr,
> LLVMConstInt(ctx->i32, swizzle, 0));
>
> value = LLVMBuildBitCast(gallivm->builder, value, ctx->i32, "");
> ac_build_indexed_store(&ctx->ac, ctx->lds,
> dw_addr, value);
> }
>
> static LLVMValueRef fetch_input_tcs(
> @@ -967,21 +966,21 @@ static LLVMValueRef fetch_input_tes(
>
> 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])
> {
> struct si_shader_context *ctx = si_shader_context(bld_base);
> - struct gallivm_state *gallivm = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> const struct tgsi_full_dst_register *reg = &inst->Dst[0];
> const struct tgsi_shader_info *sh_info = &ctx->shader->selector->info;
> unsigned chan_index;
> LLVMValueRef dw_addr, stride;
> LLVMValueRef rw_buffers, buffer, base, buf_addr;
> LLVMValueRef values[4];
> bool skip_lds_store;
> bool is_tess_factor = false;
>
> /* Only handle per-patch and per-vertex outputs here.
> @@ -1038,59 +1037,58 @@ static void store_output_tcs(struct lp_build_tgsi_context *bld_base,
> values[chan_index] = value;
>
> if (inst->Dst[0].Register.WriteMask != 0xF && !is_tess_factor) {
> ac_build_buffer_store_dword(&ctx->ac, buffer, value, 1,
> buf_addr, base,
> 4 * chan_index, 1, 0, true, false);
> }
> }
>
> if (inst->Dst[0].Register.WriteMask == 0xF && !is_tess_factor) {
> - LLVMValueRef value = lp_build_gather_values(bld_base->base.gallivm,
> + LLVMValueRef value = lp_build_gather_values(gallivm,
> values, 4);
> ac_build_buffer_store_dword(&ctx->ac, buffer, value, 4, buf_addr,
> base, 0, 1, 0, true, false);
> }
> }
>
> static LLVMValueRef fetch_input_gs(
> struct lp_build_tgsi_context *bld_base,
> const struct tgsi_full_src_register *reg,
> enum tgsi_opcode_type type,
> unsigned swizzle)
> {
> - struct lp_build_context *base = &bld_base->base;
> struct si_shader_context *ctx = si_shader_context(bld_base);
> struct si_shader *shader = ctx->shader;
> struct lp_build_context *uint = &ctx->bld_base.uint_bld;
> - struct gallivm_state *gallivm = base->gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> LLVMValueRef vtx_offset, soffset;
> unsigned vtx_offset_param;
> struct tgsi_shader_info *info = &shader->selector->info;
> unsigned semantic_name = info->input_semantic_name[reg->Register.Index];
> unsigned semantic_index = info->input_semantic_index[reg->Register.Index];
> unsigned param;
> LLVMValueRef value;
>
> if (swizzle != ~0 && semantic_name == TGSI_SEMANTIC_PRIMID)
> return get_primitive_id(bld_base, swizzle);
>
> if (!reg->Register.Dimension)
> return NULL;
>
> if (swizzle == ~0) {
> LLVMValueRef values[TGSI_NUM_CHANNELS];
> unsigned chan;
> for (chan = 0; chan < TGSI_NUM_CHANNELS; chan++) {
> values[chan] = fetch_input_gs(bld_base, reg, type, chan);
> }
> - return lp_build_gather_values(bld_base->base.gallivm, values,
> + 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;
> } else {
> assert(vtx_offset_param < 6);
> vtx_offset_param += SI_PARAM_VTX2_OFFSET - 2;
> @@ -1167,23 +1165,21 @@ static void interp_fs_input(struct si_shader_context *ctx,
> unsigned input_index,
> unsigned semantic_name,
> unsigned semantic_index,
> unsigned num_interp_inputs,
> unsigned colors_read_mask,
> LLVMValueRef interp_param,
> LLVMValueRef prim_mask,
> LLVMValueRef face,
> LLVMValueRef result[4])
> {
> - struct lp_build_tgsi_context *bld_base = &ctx->bld_base;
> - struct lp_build_context *base = &bld_base->base;
> - struct gallivm_state *gallivm = base->gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> LLVMValueRef attr_number;
> LLVMValueRef i, j;
>
> unsigned chan;
>
> /* fs.constant returns the param from the middle vertex, so it's not
> * really useful for flat shading. It's meant to be used for custom
> * interpolation (but the intrinsic can't fetch from the other two
> * vertices).
> *
> @@ -1640,21 +1636,21 @@ static LLVMValueRef fetch_constant(
>
> LLVMValueRef addr, bufp;
> LLVMValueRef result;
>
> if (swizzle == LP_CHAN_ALL) {
> unsigned chan;
> LLVMValueRef values[4];
> for (chan = 0; chan < TGSI_NUM_CHANNELS; ++chan)
> values[chan] = fetch_constant(bld_base, reg, type, chan);
>
> - return lp_build_gather_values(bld_base->base.gallivm, values, 4);
> + 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 index;
> index = get_bounded_indirect_index(ctx, ®->DimIndirect,
> reg->Dimension.Index,
> @@ -1713,21 +1709,21 @@ static LLVMValueRef si_llvm_pack_two_int32_as_int16(struct si_shader_context *ct
> }
>
> /* Initialize arguments for the shader export intrinsic */
> static void si_llvm_init_export_args(struct lp_build_tgsi_context *bld_base,
> LLVMValueRef *values,
> unsigned target,
> struct ac_export_args *args)
> {
> struct si_shader_context *ctx = si_shader_context(bld_base);
> struct lp_build_context *base = &bld_base->base;
> - LLVMBuilderRef builder = base->gallivm->builder;
> + LLVMBuilderRef builder = ctx->gallivm.builder;
> LLVMValueRef val[4];
> unsigned spi_shader_col_format = V_028714_SPI_SHADER_32_ABGR;
> unsigned chan;
> bool is_int8, is_int10;
>
> /* Default is 0xf. Adjusted below depending on the format. */
> args->enabled_channels = 0xf; /* writemask */
>
> /* Specify whether the EXEC mask represents the valid mask */
> args->valid_mask = 0;
> @@ -1783,21 +1779,21 @@ static void si_llvm_init_export_args(struct lp_build_tgsi_context *bld_base,
>
> for (chan = 0; chan < 2; chan++) {
> LLVMValueRef pack_args[2] = {
> values[2 * chan],
> values[2 * chan + 1]
> };
> LLVMValueRef packed;
>
> packed = ac_build_cvt_pkrtz_f16(&ctx->ac, pack_args);
> args->out[chan] =
> - LLVMBuildBitCast(base->gallivm->builder,
> + LLVMBuildBitCast(ctx->gallivm.builder,
> packed, ctx->f32, "");
> }
> break;
>
> case V_028714_SPI_SHADER_UNORM16_ABGR:
> for (chan = 0; chan < 4; chan++) {
> val[chan] = ac_build_clamp(&ctx->ac, values[chan]);
> val[chan] = LLVMBuildFMul(builder, val[chan],
> LLVMConstReal(ctx->f32, 65535), "");
> val[chan] = LLVMBuildFAdd(builder, val[chan],
> @@ -1922,21 +1918,21 @@ static void si_alpha_test(struct lp_build_tgsi_context *bld_base,
> } else {
> ac_build_kill(&ctx->ac, NULL);
> }
> }
>
> static LLVMValueRef si_scale_alpha_by_sample_mask(struct lp_build_tgsi_context *bld_base,
> LLVMValueRef alpha,
> unsigned samplemask_param)
> {
> struct si_shader_context *ctx = si_shader_context(bld_base);
> - struct gallivm_state *gallivm = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> LLVMValueRef coverage;
>
> /* alpha = alpha * popcount(coverage) / SI_NUM_SMOOTH_AA_SAMPLES */
> coverage = LLVMGetParam(ctx->main_fn,
> samplemask_param);
> coverage = bitcast(bld_base, TGSI_TYPE_SIGNED, coverage);
>
> coverage = lp_build_intrinsic(gallivm->builder, "llvm.ctpop.i32",
> ctx->i32,
> &coverage, 1, LP_FUNC_ATTR_READNONE);
> @@ -2300,29 +2296,29 @@ handle_semantic:
> pos_args[1].out[1] = base->zero; /* Y */
> pos_args[1].out[2] = base->zero; /* Z */
> pos_args[1].out[3] = base->zero; /* W */
>
> if (shader->selector->info.writes_psize)
> pos_args[1].out[0] = psize_value;
>
> if (shader->selector->info.writes_edgeflag) {
> /* The output is a float, but the hw expects an integer
> * with the first bit containing the edge flag. */
> - edgeflag_value = LLVMBuildFPToUI(base->gallivm->builder,
> + edgeflag_value = LLVMBuildFPToUI(ctx->gallivm.builder,
> edgeflag_value,
> ctx->i32, "");
> edgeflag_value = lp_build_min(&bld_base->int_bld,
> edgeflag_value,
> ctx->i32_1);
>
> /* The LLVM intrinsic expects a float. */
> - pos_args[1].out[1] = LLVMBuildBitCast(base->gallivm->builder,
> + pos_args[1].out[1] = LLVMBuildBitCast(ctx->gallivm.builder,
> edgeflag_value,
> ctx->f32, "");
> }
>
> if (shader->selector->info.writes_layer)
> pos_args[1].out[2] = layer_value;
>
> if (shader->selector->info.writes_viewport_index)
> pos_args[1].out[3] = viewport_index_value;
> }
> @@ -2347,21 +2343,21 @@ handle_semantic:
> }
> }
>
> /**
> * 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);
> - struct gallivm_state *gallivm = bld_base->base.gallivm;
> + 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);
>
> rw_buffers = LLVMGetParam(ctx->main_fn, SI_PARAM_RW_BUFFERS);
> buffer = ac_build_indexed_load_const(&ctx->ac, rw_buffers,
> LLVMConstInt(ctx->i32, SI_HS_RING_TESS_OFFCHIP, 0));
>
> @@ -2393,21 +2389,21 @@ static void si_copy_tcs_inputs(struct lp_build_tgsi_context *bld_base)
> buffer_offset, 0, 1, 0, true, false);
> }
> }
>
> static void si_write_tess_factors(struct lp_build_tgsi_context *bld_base,
> LLVMValueRef rel_patch_id,
> LLVMValueRef invocation_id,
> LLVMValueRef tcs_out_current_patch_data_offset)
> {
> struct si_shader_context *ctx = si_shader_context(bld_base);
> - struct gallivm_state *gallivm = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> struct si_shader *shader = ctx->shader;
> unsigned tess_inner_index, tess_outer_index;
> LLVMValueRef lds_base, lds_inner, lds_outer, byteoffset, buffer;
> LLVMValueRef out[6], vec0, vec1, rw_buffers, tf_base, inner[4], outer[4];
> unsigned stride, outer_comps, inner_comps, i;
> struct lp_build_if_state if_ctx, inner_if_ctx;
>
> si_llvm_emit_barrier(NULL, bld_base, NULL);
>
> /* Do this only for invocation 0, because the tess levels are per-patch,
> @@ -2563,21 +2559,21 @@ static void si_llvm_emit_tcs_epilogue(struct lp_build_tgsi_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);
> tf_lds_offset = get_tcs_out_current_patch_data_offset(ctx);
>
> /* Return epilog parameters from this function. */
> - LLVMBuilderRef builder = bld_base->base.gallivm->builder;
> + 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);
> rw_buffers = LLVMBuildPtrToInt(builder, rw_buffers, ctx->i64, "");
> rw_buffers = LLVMBuildBitCast(builder, rw_buffers, ctx->v2i32, "");
> rw0 = LLVMBuildExtractElement(builder, rw_buffers,
> @@ -2610,21 +2606,21 @@ static void si_llvm_emit_tcs_epilogue(struct lp_build_tgsi_context *bld_base)
> ret = LLVMBuildInsertValue(builder, ret, invocation_id, vgpr++, "");
> ret = LLVMBuildInsertValue(builder, ret, tf_lds_offset, vgpr++, "");
> ctx->return_value = ret;
> }
>
> 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 = bld_base->base.gallivm;
> + 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_LS_OUT_LAYOUT, 13, 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. */
> @@ -2639,21 +2635,21 @@ static void si_llvm_emit_ls_epilogue(struct lp_build_tgsi_context *bld_base)
> for (chan = 0; chan < 4; chan++) {
> lds_store(bld_base, chan, dw_addr,
> LLVMBuildLoad(gallivm->builder, out_ptr[chan], ""));
> }
> }
> }
>
> static void si_llvm_emit_es_epilogue(struct lp_build_tgsi_context *bld_base)
> {
> struct si_shader_context *ctx = si_shader_context(bld_base);
> - struct gallivm_state *gallivm = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> struct si_shader *es = ctx->shader;
> struct tgsi_shader_info *info = &es->selector->info;
> LLVMValueRef soffset = LLVMGetParam(ctx->main_fn,
> ctx->param_es2gs_offset);
> unsigned chan;
> int i;
>
> for (i = 0; i < info->num_outputs; i++) {
> LLVMValueRef *out_ptr = ctx->outputs[i];
> int param_index;
> @@ -2682,21 +2678,21 @@ 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));
> }
>
> 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 = bld_base->base.gallivm;
> + 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;
>
> assert(!ctx->shader->is_gs_copy_shader);
>
> outputs = MALLOC((info->num_outputs + 1) * sizeof(outputs[0]));
>
> /* Vertex color clamping.
> *
> @@ -2815,21 +2811,21 @@ static void si_export_mrt_z(struct lp_build_tgsi_context *bld_base,
> args.out[2] = base->undef; /* B, sample mask */
> args.out[3] = base->undef; /* A, alpha to mask */
>
> if (format == V_028710_SPI_SHADER_UINT16_ABGR) {
> assert(!depth);
> args.compr = 1; /* COMPR flag */
>
> if (stencil) {
> /* Stencil should be in X[23:16]. */
> stencil = bitcast(bld_base, TGSI_TYPE_UNSIGNED, stencil);
> - stencil = LLVMBuildShl(base->gallivm->builder, stencil,
> + stencil = LLVMBuildShl(ctx->gallivm.builder, stencil,
> LLVMConstInt(ctx->i32, 16, 0), "");
> args.out[0] = bitcast(bld_base, TGSI_TYPE_FLOAT, stencil);
> mask |= 0x3;
> }
> if (samplemask) {
> /* SampleMask should be in Y[15:0]. */
> args.out[1] = samplemask;
> mask |= 0xc;
> }
> } else {
> @@ -2963,23 +2959,22 @@ static void si_export_null(struct lp_build_tgsi_context *bld_base)
> * vN+1 = Stencil
> * vN+2 = SampleMask
> * vN+3 = SampleMaskIn (used for OpenGL smoothing)
> *
> * The alpha-ref SGPR is returned via its original location.
> */
> static void si_llvm_return_fs_outputs(struct lp_build_tgsi_context *bld_base)
> {
> struct si_shader_context *ctx = si_shader_context(bld_base);
> struct si_shader *shader = ctx->shader;
> - struct lp_build_context *base = &bld_base->base;
> struct tgsi_shader_info *info = &shader->selector->info;
> - LLVMBuilderRef builder = base->gallivm->builder;
> + LLVMBuilderRef builder = ctx->gallivm.builder;
> unsigned i, j, first_vgpr, vgpr;
>
> LLVMValueRef color[8][4] = {};
> LLVMValueRef depth = NULL, stencil = NULL, samplemask = NULL;
> LLVMValueRef ret;
>
> /* Read the output values. */
> for (i = 0; i < info->num_outputs; i++) {
> unsigned semantic_name = info->output_semantic_name[i];
> unsigned semantic_index = info->output_semantic_index[i];
> @@ -3049,21 +3044,21 @@ static void si_llvm_return_fs_outputs(struct lp_build_tgsi_context *bld_base)
>
> /**
> * Given a v8i32 resource descriptor for a buffer, extract the size of the
> * buffer in number of elements and return it as an i32.
> */
> static LLVMValueRef get_buffer_size(
> struct lp_build_tgsi_context *bld_base,
> LLVMValueRef descriptor)
> {
> struct si_shader_context *ctx = si_shader_context(bld_base);
> - struct gallivm_state *gallivm = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> LLVMBuilderRef builder = gallivm->builder;
> LLVMValueRef size =
> LLVMBuildExtractElement(builder, descriptor,
> LLVMConstInt(ctx->i32, 2, 0), "");
>
> if (ctx->screen->b.chip_class == VI) {
> /* On VI, 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.
> */
> @@ -3296,21 +3291,21 @@ image_fetch_rsrc(
> if (dcc_off && target != TGSI_TEXTURE_BUFFER)
> *rsrc = force_dcc_off(ctx, *rsrc);
> }
>
> static LLVMValueRef image_fetch_coords(
> struct lp_build_tgsi_context *bld_base,
> const struct tgsi_full_instruction *inst,
> unsigned src)
> {
> struct si_shader_context *ctx = si_shader_context(bld_base);
> - struct gallivm_state *gallivm = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> LLVMBuilderRef builder = gallivm->builder;
> unsigned target = inst->Memory.Texture;
> unsigned num_coords = tgsi_util_get_texture_coord_dim(target);
> LLVMValueRef coords[4];
> LLVMValueRef tmp;
> int chan;
>
> for (chan = 0; chan < num_coords; ++chan) {
> tmp = lp_build_emit_fetch(bld_base, inst, src, chan);
> tmp = LLVMBuildBitCast(builder, tmp, ctx->i32, "");
> @@ -3409,21 +3404,21 @@ static void buffer_append_args(
> i1true : i1false; /* glc */
> }
> emit_data->args[emit_data->arg_count++] = i1false; /* slc */
> }
>
> static void load_fetch_args(
> 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 = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> const struct tgsi_full_instruction * inst = emit_data->inst;
> unsigned target = inst->Memory.Texture;
> LLVMValueRef rsrc;
>
> emit_data->dst_type = ctx->v4f32;
>
> if (inst->Src[0].Register.File == TGSI_FILE_BUFFER) {
> LLVMBuilderRef builder = gallivm->builder;
> LLVMValueRef offset;
> LLVMValueRef tmp;
> @@ -3611,21 +3606,21 @@ static bool is_oneway_access_only(const struct tgsi_full_instruction *inst,
> }
> return false;
> }
>
> static void load_emit(
> 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 = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> LLVMBuilderRef builder = gallivm->builder;
> const struct tgsi_full_instruction * inst = emit_data->inst;
> const struct tgsi_shader_info *info = &ctx->shader->selector->info;
> char intrinsic_name[64];
> bool readonly_memory = false;
>
> if (inst->Src[0].Register.File == TGSI_FILE_MEMORY) {
> load_emit_memory(ctx, emit_data);
> return;
> }
> @@ -3664,21 +3659,21 @@ static void load_emit(
> emit_data->args, emit_data->arg_count,
> get_load_intr_attribs(readonly_memory));
> }
> }
>
> static void store_fetch_args(
> 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 = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> LLVMBuilderRef builder = gallivm->builder;
> const struct tgsi_full_instruction * inst = emit_data->inst;
> struct tgsi_full_src_register memory;
> LLVMValueRef chans[4];
> LLVMValueRef data;
> LLVMValueRef rsrc;
> unsigned chan;
>
> emit_data->dst_type = LLVMVoidTypeInContext(gallivm->context);
>
> @@ -3828,21 +3823,21 @@ static void store_emit_memory(
> LLVMBuildStore(builder, data, derived_ptr);
> }
> }
>
> static void store_emit(
> 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 = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> LLVMBuilderRef builder = gallivm->builder;
> const struct tgsi_full_instruction * inst = emit_data->inst;
> const struct tgsi_shader_info *info = &ctx->shader->selector->info;
> unsigned target = inst->Memory.Texture;
> char intrinsic_name[64];
> bool writeonly_memory = false;
>
> if (inst->Dst[0].Register.File == TGSI_FILE_MEMORY) {
> store_emit_memory(ctx, emit_data);
> return;
> @@ -3881,21 +3876,21 @@ static void store_emit(
> emit_data->args, emit_data->arg_count,
> get_store_intr_attribs(writeonly_memory));
> }
> }
>
> static void atomic_fetch_args(
> 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 = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> LLVMBuilderRef builder = gallivm->builder;
> const struct tgsi_full_instruction * inst = emit_data->inst;
> LLVMValueRef data1, data2;
> LLVMValueRef rsrc;
> LLVMValueRef tmp;
>
> emit_data->dst_type = ctx->f32;
>
> tmp = lp_build_emit_fetch(bld_base, inst, 2, 0);
> data1 = LLVMBuildBitCast(builder, tmp, ctx->i32, "");
> @@ -4009,21 +4004,21 @@ static void atomic_emit_memory(struct si_shader_context *ctx,
> }
> emit_data->output[emit_data->chan] = LLVMBuildBitCast(builder, result, emit_data->dst_type, "");
> }
>
> static void atomic_emit(
> 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 = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> LLVMBuilderRef builder = gallivm->builder;
> const struct tgsi_full_instruction * inst = emit_data->inst;
> char intrinsic_name[40];
> LLVMValueRef tmp;
>
> if (inst->Src[0].Register.File == TGSI_FILE_MEMORY) {
> atomic_emit_memory(ctx, emit_data);
> return;
> }
>
> @@ -4146,21 +4141,21 @@ static void resq_fetch_args(
> 0xf);
> }
> }
>
> static void resq_emit(
> 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 = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> LLVMBuilderRef builder = gallivm->builder;
> const struct tgsi_full_instruction *inst = emit_data->inst;
> LLVMValueRef out;
>
> if (inst->Src[0].Register.File == TGSI_FILE_BUFFER) {
> out = LLVMBuildExtractElement(builder, emit_data->args[0],
> LLVMConstInt(ctx->i32, 2, 0), "");
> } else if (inst->Memory.Texture == TGSI_TEXTURE_BUFFER) {
> out = get_buffer_size(bld_base, emit_data->args[0]);
> } else {
> @@ -4347,21 +4342,21 @@ static void txq_emit(const struct lp_build_tgsi_action *action,
> LLVMValueRef result = ac_build_image_opcode(&ctx->ac, &args);
>
> emit_data->output[emit_data->chan] = fix_resinfo(ctx, target, result);
> }
>
> static void tex_fetch_args(
> 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 = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> const struct tgsi_full_instruction *inst = emit_data->inst;
> unsigned opcode = inst->Instruction.Opcode;
> unsigned target = inst->Texture.Texture;
> LLVMValueRef coords[5], derivs[6];
> LLVMValueRef address[16];
> unsigned num_coords = tgsi_util_get_texture_coord_dim(target);
> int ref_pos = tgsi_util_get_shadow_ref_src_index(target);
> unsigned count = 0;
> unsigned chan;
> unsigned num_deriv_channels = 0;
> @@ -4873,21 +4868,21 @@ static void build_tex_intrinsic(const struct lp_build_tgsi_action *action,
> emit_data->output[emit_data->chan] =
> ac_build_image_opcode(&ctx->ac, &args);
> }
>
> static void si_llvm_emit_txqs(
> 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 = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> LLVMBuilderRef builder = gallivm->builder;
> LLVMValueRef res, samples;
> LLVMValueRef res_ptr, samp_ptr, fmask_ptr = NULL;
>
> tex_fetch_ptrs(bld_base, emit_data, &res_ptr, &samp_ptr, &fmask_ptr);
>
>
> /* Read the samples from the descriptor directly. */
> res = LLVMBuildBitCast(builder, res_ptr, ctx->v8i32, "");
> samples = LLVMBuildExtractElement(
> @@ -4902,21 +4897,21 @@ static void si_llvm_emit_txqs(
>
> emit_data->output[emit_data->chan] = samples;
> }
>
> static void si_llvm_emit_ddxy(
> 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 = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> unsigned opcode = emit_data->info->opcode;
> LLVMValueRef val;
> int idx;
> unsigned mask;
>
> if (opcode == TGSI_OPCODE_DDX_FINE)
> mask = AC_TID_MASK_LEFT;
> else if (opcode == TGSI_OPCODE_DDY_FINE)
> mask = AC_TID_MASK_TOP;
> else
> @@ -4934,40 +4929,40 @@ static void si_llvm_emit_ddxy(
> /*
> * this takes an I,J coordinate pair,
> * and works out the X and Y derivatives.
> * it returns DDX(I), DDX(J), DDY(I), DDY(J).
> */
> static LLVMValueRef si_llvm_emit_ddxy_interp(
> struct lp_build_tgsi_context *bld_base,
> LLVMValueRef interp_ij)
> {
> struct si_shader_context *ctx = si_shader_context(bld_base);
> - struct gallivm_state *gallivm = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> LLVMValueRef result[4], a;
> unsigned i;
>
> for (i = 0; i < 2; i++) {
> a = LLVMBuildExtractElement(gallivm->builder, interp_ij,
> LLVMConstInt(ctx->i32, i, 0), "");
> result[i] = lp_build_emit_llvm_unary(bld_base, TGSI_OPCODE_DDX, a);
> result[2+i] = lp_build_emit_llvm_unary(bld_base, TGSI_OPCODE_DDY, a);
> }
>
> return lp_build_gather_values(gallivm, result, 4);
> }
>
> static void interp_fetch_args(
> 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 = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> const struct tgsi_full_instruction *inst = emit_data->inst;
>
> if (inst->Instruction.Opcode == TGSI_OPCODE_INTERP_OFFSET) {
> /* offset is in second src, first two channels */
> emit_data->args[0] = lp_build_emit_fetch(bld_base,
> emit_data->inst, 1,
> TGSI_CHAN_X);
> emit_data->args[1] = lp_build_emit_fetch(bld_base,
> emit_data->inst, 1,
> TGSI_CHAN_Y);
> @@ -4998,21 +4993,21 @@ static void interp_fetch_args(
> emit_data->arg_count = 2;
> }
> }
>
> static void build_interp_intrinsic(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 si_shader *shader = ctx->shader;
> - struct gallivm_state *gallivm = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> LLVMValueRef interp_param;
> const struct tgsi_full_instruction *inst = emit_data->inst;
> int input_index = inst->Src[0].Register.Index;
> int chan;
> int i;
> LLVMValueRef attr_number;
> LLVMValueRef params = LLVMGetParam(ctx->main_fn, SI_PARAM_PRIM_MASK);
> int interp_param_idx;
> unsigned interp = shader->selector->info.input_interpolate[input_index];
> unsigned location;
> @@ -5063,21 +5058,21 @@ static void build_interp_intrinsic(const struct lp_build_tgsi_action *action,
> ctx->f32, "");
>
> temp1 = LLVMBuildFMul(gallivm->builder, ddx_el, emit_data->args[0], "");
>
> temp1 = LLVMBuildFAdd(gallivm->builder, temp1, interp_el, "");
>
> temp2 = LLVMBuildFMul(gallivm->builder, ddy_el, emit_data->args[1], "");
>
> ij_out[i] = LLVMBuildFAdd(gallivm->builder, temp2, temp1, "");
> }
> - interp_param = lp_build_gather_values(bld_base->base.gallivm, ij_out, 2);
> + interp_param = lp_build_gather_values(gallivm, ij_out, 2);
> }
>
> for (chan = 0; chan < 4; chan++) {
> LLVMValueRef llvm_chan;
> unsigned schan;
>
> schan = tgsi_util_get_full_src_register_swizzle(&inst->Src[0], chan);
> llvm_chan = LLVMConstInt(ctx->i32, schan, 0);
>
> if (interp_param) {
> @@ -5194,21 +5189,21 @@ static unsigned si_llvm_get_stream(struct lp_build_tgsi_context *bld_base,
> /* Emit one vertex from the geometry shader */
> static void si_llvm_emit_vertex(
> 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 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 = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> struct lp_build_if_state if_state;
> LLVMValueRef soffset = LLVMGetParam(ctx->main_fn,
> SI_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);
> @@ -5294,21 +5289,21 @@ static void si_llvm_emit_primitive(
> 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));
> }
>
> 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 = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
>
> /* SI 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 (HAVE_LLVM >= 0x0309 &&
> ctx->screen->b.chip_class == SI &&
> ctx->type == PIPE_SHADER_TESS_CTRL) {
> emit_waitcnt(ctx, LGKM_CNT & VM_CNT);
> return;
> @@ -5453,21 +5448,21 @@ static unsigned si_get_max_workgroup_size(struct si_shader *shader)
> * compile it for the maximum possible group size.
> */
> max_work_group_size = SI_MAX_VARIABLE_THREADS_PER_BLOCK;
> }
> 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 = bld_base->base.gallivm;
> + 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 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);
> @@ -5718,21 +5713,21 @@ static void create_function(struct si_shader_context *ctx)
> ctx->type == PIPE_SHADER_TESS_CTRL)
> declare_tess_lds(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->bld_base.base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> LLVMBuilderRef builder = gallivm->builder;
>
> LLVMValueRef buf_ptr = LLVMGetParam(ctx->main_fn,
> SI_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) {
> @@ -5820,22 +5815,21 @@ static void preload_ring_buffers(struct si_shader_context *ctx)
>
> ctx->gsvs_ring[stream] = ring;
> }
> }
> }
>
> static void si_llvm_emit_polygon_stipple(struct si_shader_context *ctx,
> LLVMValueRef param_rw_buffers,
> unsigned param_pos_fixed_pt)
> {
> - struct lp_build_tgsi_context *bld_base = &ctx->bld_base;
> - struct gallivm_state *gallivm = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> LLVMBuilderRef builder = gallivm->builder;
> LLVMValueRef slot, desc, offset, row, bit, address[2];
>
> /* Use the fixed-point gl_FragCoord input.
> * Since the stipple pattern is 32x32 and it repeats, just get 5 bits
> * per coordinate to get the repeating effect.
> */
> address[0] = unpack_param(ctx, param_pos_fixed_pt, 0, 5);
> address[1] = unpack_param(ctx, param_pos_fixed_pt, 16, 5);
>
> @@ -6436,28 +6430,28 @@ si_generate_gs_copy_shader(struct si_screen *sscreen,
> LLVMBuildBr(builder, end_bb);
> }
>
> LLVMPositionBuilderAtEnd(builder, end_bb);
>
> LLVMBuildRetVoid(gallivm->builder);
>
> /* Dump LLVM IR before any optimization passes */
> if (sscreen->b.debug_flags & DBG_PREOPT_IR &&
> r600_can_dump_shader(&sscreen->b, PIPE_SHADER_GEOMETRY))
> - ac_dump_module(bld_base->base.gallivm->module);
> + ac_dump_module(ctx.gallivm.module);
>
> si_llvm_finalize_module(&ctx,
> r600_extra_shader_checks(&sscreen->b, PIPE_SHADER_GEOMETRY));
>
> r = si_compile_llvm(sscreen, &ctx.shader->binary,
> &ctx.shader->config, ctx.tm,
> - bld_base->base.gallivm->module,
> + ctx.gallivm.module,
> debug, PIPE_SHADER_GEOMETRY,
> "GS Copy Shader");
> if (!r) {
> if (r600_can_dump_shader(&sscreen->b, PIPE_SHADER_GEOMETRY))
> fprintf(stderr, "GS Copy Shader:\n");
> si_shader_dump(sscreen, ctx.shader, debug,
> PIPE_SHADER_GEOMETRY, stderr, true);
> r = si_shader_binary_upload(sscreen, ctx.shader);
> }
>
> @@ -6859,21 +6853,21 @@ static bool si_compile_tgsi_main(struct si_shader_context *ctx,
> return false;
> }
>
> create_function(ctx);
> preload_ring_buffers(ctx);
>
> if (ctx->type == PIPE_SHADER_GEOMETRY) {
> int i;
> for (i = 0; i < 4; i++) {
> ctx->gs_next_vertex[i] =
> - lp_build_alloca(bld_base->base.gallivm,
> + lp_build_alloca(&ctx->gallivm,
> ctx->i32, "");
> }
> }
>
> if (!lp_build_tgsi_llvm(bld_base, sel->tokens)) {
> fprintf(stderr, "Failed to translate shader from TGSI to LLVM\n");
> return false;
> }
>
> si_llvm_build_ret(ctx, ctx->return_value);
> @@ -7339,41 +7333,39 @@ static void si_build_wrapper_function(struct si_shader_context *ctx,
> }
>
> int si_compile_tgsi_shader(struct si_screen *sscreen,
> LLVMTargetMachineRef tm,
> struct si_shader *shader,
> bool is_monolithic,
> struct pipe_debug_callback *debug)
> {
> struct si_shader_selector *sel = shader->selector;
> struct si_shader_context ctx;
> - struct lp_build_tgsi_context *bld_base;
> LLVMModuleRef mod;
> int r = -1;
>
> /* Dump TGSI code before doing TGSI->LLVM conversion in case the
> * conversion fails. */
> if (r600_can_dump_shader(&sscreen->b, sel->info.processor) &&
> !(sscreen->b.debug_flags & DBG_NO_TGSI)) {
> tgsi_dump(sel->tokens, 0);
> si_dump_streamout(&sel->so);
> }
>
> si_init_shader_ctx(&ctx, sscreen, shader, tm);
> ctx.separate_prolog = !is_monolithic;
>
> memset(shader->info.vs_output_param_offset, EXP_PARAM_UNDEFINED,
> sizeof(shader->info.vs_output_param_offset));
>
> shader->info.uses_instanceid = sel->info.uses_instanceid;
>
> - bld_base = &ctx.bld_base;
> ctx.load_system_value = declare_system_value;
>
> if (!si_compile_tgsi_main(&ctx, shader)) {
> si_llvm_dispose(&ctx);
> return -1;
> }
>
> if (is_monolithic && ctx.type == PIPE_SHADER_VERTEX) {
> LLVMValueRef parts[3];
> bool need_prolog;
> @@ -7452,21 +7444,21 @@ int si_compile_tgsi_shader(struct si_screen *sscreen,
> parts[0] = ctx.main_fn;
> }
>
> si_get_ps_epilog_key(shader, &epilog_key);
> si_build_ps_epilog_function(&ctx, &epilog_key);
> parts[need_prolog ? 2 : 1] = ctx.main_fn;
>
> si_build_wrapper_function(&ctx, parts, need_prolog ? 3 : 2, need_prolog ? 1 : 0);
> }
>
> - mod = bld_base->base.gallivm->module;
> + mod = ctx.gallivm.module;
>
> /* Dump LLVM IR before any optimization passes */
> if (sscreen->b.debug_flags & DBG_PREOPT_IR &&
> r600_can_dump_shader(&sscreen->b, ctx.type))
> ac_dump_module(mod);
>
> si_llvm_finalize_module(&ctx,
> r600_extra_shader_checks(&sscreen->b, ctx.type));
>
> /* Post-optimization transformations and analysis. */
> diff --git a/src/gallium/drivers/radeonsi/si_shader_tgsi_alu.c b/src/gallium/drivers/radeonsi/si_shader_tgsi_alu.c
> index d7ec9ec..1e2d75d 100644
> --- a/src/gallium/drivers/radeonsi/si_shader_tgsi_alu.c
> +++ b/src/gallium/drivers/radeonsi/si_shader_tgsi_alu.c
> @@ -494,21 +494,21 @@ static void emit_bfi(const struct lp_build_tgsi_action *action,
> lp_build_const_int32(gallivm, 32), "");
> emit_data->output[emit_data->chan] =
> LLVMBuildSelect(builder, cond, emit_data->args[1], bfi_sm5, "");
> }
>
> static void emit_bfe(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 = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> LLVMBuilderRef builder = gallivm->builder;
> LLVMValueRef bfe_sm5;
> LLVMValueRef cond;
>
> bfe_sm5 = ac_build_bfe(&ctx->ac, emit_data->args[0],
> emit_data->args[1], emit_data->args[2],
> emit_data->info->opcode == TGSI_OPCODE_IBFE);
>
> /* Correct for GLSL semantics. */
> cond = LLVMBuildICmp(builder, LLVMIntUGE, emit_data->args[2],
> @@ -690,21 +690,21 @@ static void emit_up2h(const struct lp_build_tgsi_action *action,
> }
> }
>
> static void emit_fdiv(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);
>
> emit_data->output[emit_data->chan] =
> - LLVMBuildFDiv(bld_base->base.gallivm->builder,
> + LLVMBuildFDiv(ctx->gallivm.builder,
> emit_data->args[0], emit_data->args[1], "");
>
> /* Use v_rcp_f32 instead of precise division. */
> if (HAVE_LLVM >= 0x0309 &&
> !LLVMIsConstant(emit_data->output[emit_data->chan]))
> LLVMSetMetadata(emit_data->output[emit_data->chan],
> ctx->fpmath_md_kind, ctx->fpmath_md_2p5_ulp);
> }
>
> /* 1/sqrt is translated to rsq for f32 if fp32 denormals are not enabled in
> diff --git a/src/gallium/drivers/radeonsi/si_shader_tgsi_setup.c b/src/gallium/drivers/radeonsi/si_shader_tgsi_setup.c
> index 341c18d..3e38f0d 100644
> --- a/src/gallium/drivers/radeonsi/si_shader_tgsi_setup.c
> +++ b/src/gallium/drivers/radeonsi/si_shader_tgsi_setup.c
> @@ -418,21 +418,21 @@ get_array_range(struct lp_build_tgsi_context *bld_base,
> range.First = 0;
> range.Last = bld_base->info->file_max[File];
> return range;
> }
>
> static LLVMValueRef
> emit_array_index(struct si_shader_context *ctx,
> const struct tgsi_ind_register *reg,
> unsigned offset)
> {
> - struct gallivm_state *gallivm = ctx->bld_base.base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
>
> if (!reg) {
> return LLVMConstInt(ctx->i32, offset, 0);
> }
> LLVMValueRef addr = LLVMBuildLoad(gallivm->builder, ctx->addrs[reg->Index][reg->Swizzle], "");
> return LLVMBuildAdd(gallivm->builder, addr, LLVMConstInt(ctx->i32, offset, 0), "");
> }
>
> /**
> * For indirect registers, construct a pointer directly to the requested
> @@ -443,21 +443,21 @@ emit_array_index(struct si_shader_context *ctx,
> */
> static LLVMValueRef
> get_pointer_into_array(struct si_shader_context *ctx,
> unsigned file,
> unsigned swizzle,
> unsigned reg_index,
> const struct tgsi_ind_register *reg_indirect)
> {
> unsigned array_id;
> struct tgsi_array_info *array;
> - struct gallivm_state *gallivm = ctx->bld_base.base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> LLVMBuilderRef builder = gallivm->builder;
> LLVMValueRef idxs[2];
> LLVMValueRef index;
> LLVMValueRef alloca;
>
> if (file != TGSI_FILE_TEMPORARY)
> return NULL;
>
> array_id = get_temp_array_id(&ctx->bld_base, reg_index, reg_indirect);
> if (!array_id)
> @@ -526,21 +526,21 @@ si_llvm_emit_fetch_64bit(struct lp_build_tgsi_context *bld_base,
> }
>
> static LLVMValueRef
> emit_array_fetch(struct lp_build_tgsi_context *bld_base,
> unsigned File, enum tgsi_opcode_type type,
> struct tgsi_declaration_range range,
> unsigned swizzle)
> {
> struct si_shader_context *ctx = si_shader_context(bld_base);
>
> - LLVMBuilderRef builder = bld_base->base.gallivm->builder;
> + LLVMBuilderRef builder = ctx->gallivm.builder;
>
> unsigned i, size = range.Last - range.First + 1;
> LLVMTypeRef vec = LLVMVectorType(tgsi2llvmtype(bld_base, type), size);
> LLVMValueRef result = LLVMGetUndef(vec);
>
> struct tgsi_full_src_register tmp_reg = {};
> tmp_reg.Register.File = File;
>
> for (i = 0; i < size; ++i) {
> tmp_reg.Register.Index = i + range.First;
> @@ -553,21 +553,21 @@ emit_array_fetch(struct lp_build_tgsi_context *bld_base,
>
> static LLVMValueRef
> load_value_from_array(struct lp_build_tgsi_context *bld_base,
> unsigned file,
> enum tgsi_opcode_type type,
> unsigned swizzle,
> unsigned reg_index,
> const struct tgsi_ind_register *reg_indirect)
> {
> struct si_shader_context *ctx = si_shader_context(bld_base);
> - struct gallivm_state *gallivm = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> LLVMBuilderRef builder = gallivm->builder;
> LLVMValueRef ptr;
>
> ptr = get_pointer_into_array(ctx, file, swizzle, reg_index, reg_indirect);
> if (ptr) {
> LLVMValueRef val = LLVMBuildLoad(builder, ptr, "");
> if (tgsi_type_is_64bit(type)) {
> LLVMValueRef ptr_hi, val_hi;
> ptr_hi = LLVMBuildGEP(builder, ptr, &ctx->i32_1, 1, "");
> val_hi = LLVMBuildLoad(builder, ptr_hi, "");
> @@ -588,21 +588,21 @@ load_value_from_array(struct lp_build_tgsi_context *bld_base,
>
> static void
> store_value_to_array(struct lp_build_tgsi_context *bld_base,
> LLVMValueRef value,
> unsigned file,
> unsigned chan_index,
> unsigned reg_index,
> const struct tgsi_ind_register *reg_indirect)
> {
> struct si_shader_context *ctx = si_shader_context(bld_base);
> - struct gallivm_state *gallivm = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> LLVMBuilderRef builder = gallivm->builder;
> LLVMValueRef ptr;
>
> ptr = get_pointer_into_array(ctx, file, chan_index, reg_index, reg_indirect);
> if (ptr) {
> LLVMBuildStore(builder, value, ptr);
> } else {
> unsigned i, size;
> struct tgsi_declaration_range range = get_array_range(bld_base, file, reg_index, reg_indirect);
> LLVMValueRef index = emit_array_index(ctx, reg_indirect, reg_index - range.First);
> @@ -657,30 +657,30 @@ get_output_ptr(struct lp_build_tgsi_context *bld_base, unsigned index,
> assert(index <= ctx->bld_base.info->file_max[TGSI_FILE_OUTPUT]);
> return ctx->outputs[index][chan];
> }
>
> LLVMValueRef si_llvm_emit_fetch(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);
> - LLVMBuilderRef builder = bld_base->base.gallivm->builder;
> + LLVMBuilderRef builder = ctx->gallivm.builder;
> LLVMValueRef result = NULL, ptr, ptr2;
>
> if (swizzle == ~0) {
> LLVMValueRef values[TGSI_NUM_CHANNELS];
> unsigned chan;
> for (chan = 0; chan < TGSI_NUM_CHANNELS; chan++) {
> values[chan] = si_llvm_emit_fetch(bld_base, reg, type, chan);
> }
> - return lp_build_gather_values(bld_base->base.gallivm, values,
> + return lp_build_gather_values(&ctx->gallivm, values,
> TGSI_NUM_CHANNELS);
> }
>
> if (reg->Register.Indirect) {
> LLVMValueRef load = load_value_from_array(bld_base, reg->Register.File, type,
> swizzle, reg->Register.Index, ®->Indirect);
> return bitcast(bld_base, type, load);
> }
>
> switch(reg->Register.File) {
> @@ -755,35 +755,35 @@ LLVMValueRef si_llvm_emit_fetch(struct lp_build_tgsi_context *bld_base,
>
> return bitcast(bld_base, type, result);
> }
>
> static LLVMValueRef fetch_system_value(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);
> - struct gallivm_state *gallivm = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
>
> LLVMValueRef cval = ctx->system_values[reg->Register.Index];
> if (LLVMGetTypeKind(LLVMTypeOf(cval)) == LLVMVectorTypeKind) {
> cval = LLVMBuildExtractElement(gallivm->builder, cval,
> LLVMConstInt(ctx->i32, swizzle, 0), "");
> }
> return bitcast(bld_base, type, cval);
> }
>
> static void emit_declaration(struct lp_build_tgsi_context *bld_base,
> const struct tgsi_full_declaration *decl)
> {
> struct si_shader_context *ctx = si_shader_context(bld_base);
> - LLVMBuilderRef builder = bld_base->base.gallivm->builder;
> + LLVMBuilderRef builder = ctx->gallivm.builder;
> unsigned first, last, i;
> switch(decl->Declaration.File) {
> case TGSI_FILE_ADDRESS:
> {
> unsigned idx;
> for (idx = decl->Range.First; idx <= decl->Range.Last; idx++) {
> unsigned chan;
> for (chan = 0; chan < TGSI_NUM_CHANNELS; chan++) {
> ctx->addrs[idx][chan] = lp_build_alloca_undef(
> &ctx->gallivm,
> @@ -846,40 +846,40 @@ static void emit_declaration(struct lp_build_tgsi_context *bld_base,
> ctx->temps_count = bld_base->info->file_max[TGSI_FILE_TEMPORARY] + 1;
> ctx->temps = MALLOC(TGSI_NUM_CHANNELS * ctx->temps_count * sizeof(LLVMValueRef));
> }
> if (!array_alloca) {
> for (i = 0; i < decl_size; ++i) {
> #ifdef DEBUG
> snprintf(name, sizeof(name), "TEMP%d.%c",
> first + i / 4, "xyzw"[i % 4]);
> #endif
> ctx->temps[first * TGSI_NUM_CHANNELS + i] =
> - lp_build_alloca_undef(bld_base->base.gallivm,
> + lp_build_alloca_undef(&ctx->gallivm,
> ctx->f32,
> name);
> }
> } else {
> LLVMValueRef idxs[2] = {
> ctx->i32_0,
> NULL
> };
> unsigned j = 0;
>
> if (writemask != TGSI_WRITEMASK_XYZW &&
> !ctx->undef_alloca) {
> /* Create a dummy alloca. We use it so that we
> * have a pointer that is safe to load from if
> * a shader ever reads from a channel that
> * it never writes to.
> */
> ctx->undef_alloca = lp_build_alloca_undef(
> - bld_base->base.gallivm,
> + &ctx->gallivm,
> ctx->f32, "undef");
> }
>
> for (i = 0; i < decl_size; ++i) {
> LLVMValueRef ptr;
> if (writemask & (1 << (i % 4))) {
> #ifdef DEBUG
> snprintf(name, sizeof(name), "TEMP%d.%c",
> first + i / 4, "xyzw"[i % 4]);
> #endif
> @@ -953,23 +953,23 @@ static void emit_declaration(struct lp_build_tgsi_context *bld_base,
> break;
> }
> }
>
> void si_llvm_emit_store(struct lp_build_tgsi_context *bld_base,
> const struct tgsi_full_instruction *inst,
> const struct tgsi_opcode_info *info,
> LLVMValueRef dst[4])
> {
> struct si_shader_context *ctx = si_shader_context(bld_base);
> - struct gallivm_state *gallivm = ctx->bld_base.base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> const struct tgsi_full_dst_register *reg = &inst->Dst[0];
> - LLVMBuilderRef builder = ctx->bld_base.base.gallivm->builder;
> + LLVMBuilderRef builder = ctx->gallivm.builder;
> LLVMValueRef temp_ptr, temp_ptr2 = NULL;
> unsigned chan, chan_index;
> bool is_vec_store = false;
> enum tgsi_opcode_type dtype = tgsi_opcode_infer_dst_type(inst->Instruction.Opcode);
>
> if (dst[0]) {
> LLVMTypeKind k = LLVMGetTypeKind(LLVMTypeOf(dst[0]));
> is_vec_store = (k == LLVMVectorTypeKind);
> }
>
> @@ -1084,112 +1084,112 @@ static void emit_default_branch(LLVMBuilderRef builder, LLVMBasicBlockRef target
> {
> if (!LLVMGetBasicBlockTerminator(LLVMGetInsertBlock(builder)))
> LLVMBuildBr(builder, target);
> }
>
> static void bgnloop_emit(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 = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> struct si_llvm_flow *flow = push_flow(ctx);
> flow->loop_entry_block = append_basic_block(ctx, "LOOP");
> flow->next_block = append_basic_block(ctx, "ENDLOOP");
> set_basicblock_name(flow->loop_entry_block, "loop", bld_base->pc);
> LLVMBuildBr(gallivm->builder, flow->loop_entry_block);
> LLVMPositionBuilderAtEnd(gallivm->builder, flow->loop_entry_block);
> }
>
> static void brk_emit(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 = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> struct si_llvm_flow *flow = get_innermost_loop(ctx);
>
> LLVMBuildBr(gallivm->builder, flow->next_block);
> }
>
> static void cont_emit(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 = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> struct si_llvm_flow *flow = get_innermost_loop(ctx);
>
> LLVMBuildBr(gallivm->builder, flow->loop_entry_block);
> }
>
> static void else_emit(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 = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> struct si_llvm_flow *current_branch = get_current_flow(ctx);
> LLVMBasicBlockRef endif_block;
>
> assert(!current_branch->loop_entry_block);
>
> endif_block = append_basic_block(ctx, "ENDIF");
> emit_default_branch(gallivm->builder, endif_block);
>
> LLVMPositionBuilderAtEnd(gallivm->builder, current_branch->next_block);
> set_basicblock_name(current_branch->next_block, "else", bld_base->pc);
>
> current_branch->next_block = endif_block;
> }
>
> static void endif_emit(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 = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> struct si_llvm_flow *current_branch = get_current_flow(ctx);
>
> assert(!current_branch->loop_entry_block);
>
> emit_default_branch(gallivm->builder, current_branch->next_block);
> LLVMPositionBuilderAtEnd(gallivm->builder, current_branch->next_block);
> set_basicblock_name(current_branch->next_block, "endif", bld_base->pc);
>
> ctx->flow_depth--;
> }
>
> static void endloop_emit(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 = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> struct si_llvm_flow *current_loop = get_current_flow(ctx);
>
> assert(current_loop->loop_entry_block);
>
> emit_default_branch(gallivm->builder, current_loop->loop_entry_block);
>
> LLVMPositionBuilderAtEnd(gallivm->builder, current_loop->next_block);
> set_basicblock_name(current_loop->next_block, "endloop", bld_base->pc);
> ctx->flow_depth--;
> }
>
> static void if_cond_emit(const struct lp_build_tgsi_action *action,
> struct lp_build_tgsi_context *bld_base,
> struct lp_build_emit_data *emit_data,
> LLVMValueRef cond)
> {
> struct si_shader_context *ctx = si_shader_context(bld_base);
> - struct gallivm_state *gallivm = bld_base->base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> struct si_llvm_flow *flow = push_flow(ctx);
> LLVMBasicBlockRef if_block;
>
> if_block = append_basic_block(ctx, "IF");
> flow->next_block = append_basic_block(ctx, "ELSE");
> set_basicblock_name(if_block, "if", bld_base->pc);
> LLVMBuildCondBr(gallivm->builder, cond, if_block, flow->next_block);
> LLVMPositionBuilderAtEnd(gallivm->builder, if_block);
> }
>
> @@ -1382,21 +1382,21 @@ void si_llvm_create_func(struct si_shader_context *ctx,
> main_fn_type = LLVMFunctionType(ret_type, ParamTypes, ParamCount, 0);
> ctx->main_fn = LLVMAddFunction(ctx->gallivm.module, name, main_fn_type);
> main_fn_body = LLVMAppendBasicBlockInContext(ctx->gallivm.context,
> ctx->main_fn, "main_body");
> LLVMPositionBuilderAtEnd(ctx->gallivm.builder, main_fn_body);
> }
>
> void si_llvm_finalize_module(struct si_shader_context *ctx,
> bool run_verifier)
> {
> - struct gallivm_state *gallivm = ctx->bld_base.base.gallivm;
> + struct gallivm_state *gallivm = &ctx->gallivm;
> const char *triple = LLVMGetTarget(gallivm->module);
> LLVMTargetLibraryInfoRef target_library_info;
>
> /* Create the pass manager */
> gallivm->passmgr = LLVMCreatePassManager();
>
> target_library_info = gallivm_create_target_library_info(triple);
> LLVMAddTargetLibraryInfo(target_library_info, gallivm->passmgr);
>
> if (run_verifier)
> @@ -1417,22 +1417,22 @@ void si_llvm_finalize_module(struct si_shader_context *ctx,
> /* Run the pass */
> LLVMRunPassManager(gallivm->passmgr, ctx->gallivm.module);
>
> LLVMDisposeBuilder(gallivm->builder);
> LLVMDisposePassManager(gallivm->passmgr);
> gallivm_dispose_target_library_info(target_library_info);
> }
>
> void si_llvm_dispose(struct si_shader_context *ctx)
> {
> - LLVMDisposeModule(ctx->bld_base.base.gallivm->module);
> - LLVMContextDispose(ctx->bld_base.base.gallivm->context);
> + LLVMDisposeModule(ctx->gallivm.module);
> + LLVMContextDispose(ctx->gallivm.context);
> FREE(ctx->temp_arrays);
> ctx->temp_arrays = NULL;
> FREE(ctx->temp_array_allocas);
> ctx->temp_array_allocas = NULL;
> FREE(ctx->temps);
> ctx->temps = NULL;
> ctx->temps_count = 0;
> FREE(ctx->imms);
> ctx->imms = NULL;
> ctx->imms_num = 0;
>
--
Lerne, wie die Welt wirklich ist,
Aber vergiss niemals, wie sie sein sollte.
More information about the mesa-dev
mailing list