From: Marek Olšák <marek.ol...@amd.com>
---
src/gallium/drivers/radeonsi/si_shader.c | 1 -
src/gallium/drivers/radeonsi/si_shader_internal.h | 1 -
.../drivers/radeonsi/si_shader_tgsi_setup.c | 53 +++++++++-------------
3 files changed, 22 insertions(+), 33 deletions(-)
diff --git a/src/gallium/drivers/radeonsi/si_shader.c
b/src/gallium/drivers/radeonsi/si_shader.c
index 8bdde1a..fed8639 100644
--- a/src/gallium/drivers/radeonsi/si_shader.c
+++ b/src/gallium/drivers/radeonsi/si_shader.c
@@ -5676,21 +5676,20 @@ static const struct lp_build_tgsi_action interp_action
= {
static void si_create_function(struct si_shader_context *ctx,
const char *name,
LLVMTypeRef *returns, unsigned num_returns,
LLVMTypeRef *params, unsigned num_params,
int last_sgpr, unsigned max_workgroup_size)
{
int i;
si_llvm_create_func(ctx, name, returns, num_returns,
params, num_params);
- si_llvm_shader_type(ctx->main_fn, ctx->type);
ctx->return_value = LLVMGetUndef(ctx->return_type);
for (i = 0; i <= last_sgpr; ++i) {
LLVMValueRef P = LLVMGetParam(ctx->main_fn, i);
/* The combination of:
* - ByVal
* - dereferenceable
* - invariant.load
* allows the optimization passes to move loads and reduces
diff --git a/src/gallium/drivers/radeonsi/si_shader_internal.h
b/src/gallium/drivers/radeonsi/si_shader_internal.h
index b54db20..35315ca 100644
--- a/src/gallium/drivers/radeonsi/si_shader_internal.h
+++ b/src/gallium/drivers/radeonsi/si_shader_internal.h
@@ -233,21 +233,20 @@ struct si_shader_context {
LLVMValueRef shared_memory;
};
static inline struct si_shader_context *
si_shader_context(struct lp_build_tgsi_context *bld_base)
{
return (struct si_shader_context*)bld_base;
}
void si_llvm_add_attribute(LLVMValueRef F, const char *name, int value);
-void si_llvm_shader_type(LLVMValueRef F, unsigned type);
LLVMTargetRef si_llvm_get_amdgpu_target(const char *triple);
unsigned si_llvm_compile(LLVMModuleRef M, struct ac_shader_binary *binary,
LLVMTargetMachineRef tm,
struct pipe_debug_callback *debug);
LLVMTypeRef tgsi2llvmtype(struct lp_build_tgsi_context *bld_base,
enum tgsi_opcode_type type);
diff --git a/src/gallium/drivers/radeonsi/si_shader_tgsi_setup.c
b/src/gallium/drivers/radeonsi/si_shader_tgsi_setup.c
index 2b0d600..de671ef 100644
--- a/src/gallium/drivers/radeonsi/si_shader_tgsi_setup.c
+++ b/src/gallium/drivers/radeonsi/si_shader_tgsi_setup.c
@@ -58,51 +58,20 @@ enum si_llvm_calling_convention {
};
void si_llvm_add_attribute(LLVMValueRef F, const char *name, int value)
{
char str[16];
snprintf(str, sizeof(str), "%i", value);
LLVMAddTargetDependentFunctionAttr(F, name, str);
}
-/**
- * Set the shader type we want to compile
- *
- * @param type shader type to set
- */
-void si_llvm_shader_type(LLVMValueRef F, unsigned type)
-{
- enum si_llvm_calling_convention calling_conv;
-
- switch (type) {
- case PIPE_SHADER_VERTEX:
- case PIPE_SHADER_TESS_CTRL:
- case PIPE_SHADER_TESS_EVAL:
- calling_conv = RADEON_LLVM_AMDGPU_VS;
- break;
- case PIPE_SHADER_GEOMETRY:
- calling_conv = RADEON_LLVM_AMDGPU_GS;
- break;
- case PIPE_SHADER_FRAGMENT:
- calling_conv = RADEON_LLVM_AMDGPU_PS;
- break;
- case PIPE_SHADER_COMPUTE:
- calling_conv = RADEON_LLVM_AMDGPU_CS;
- break;
- default:
- unreachable("Unhandle shader type");
- }
-
- LLVMSetFunctionCallConv(F, calling_conv);
-}
-
static void init_amdgpu_target()
{
gallivm_init_llvm_targets();
LLVMInitializeAMDGPUTargetInfo();
LLVMInitializeAMDGPUTarget();
LLVMInitializeAMDGPUTargetMC();
LLVMInitializeAMDGPUAsmPrinter();
/* For inline assembly. */
LLVMInitializeAMDGPUAsmParser();
@@ -1385,35 +1354,57 @@ void si_llvm_context_set_tgsi(struct si_shader_context
*ctx,
ctx->bld_base.emit_fetch_funcs[TGSI_FILE_SYSTEM_VALUE] =
fetch_system_value;
}
void si_llvm_create_func(struct si_shader_context *ctx,
const char *name,
LLVMTypeRef *return_types, unsigned num_return_elems,
LLVMTypeRef *ParamTypes, unsigned ParamCount)
{
LLVMTypeRef main_fn_type, ret_type;
LLVMBasicBlockRef main_fn_body;
+ enum si_llvm_calling_convention call_conv;
if (num_return_elems)
ret_type = LLVMStructTypeInContext(ctx->gallivm.context,
return_types,
num_return_elems, true);
else
ret_type = LLVMVoidTypeInContext(ctx->gallivm.context);
/* Setup the function */
ctx->return_type = ret_type;
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);
+
+ switch (ctx->type) {
+ case PIPE_SHADER_VERTEX:
+ case PIPE_SHADER_TESS_CTRL:
+ case PIPE_SHADER_TESS_EVAL:
+ call_conv = RADEON_LLVM_AMDGPU_VS;
+ break;
+ case PIPE_SHADER_GEOMETRY:
+ call_conv = RADEON_LLVM_AMDGPU_GS;
+ break;
+ case PIPE_SHADER_FRAGMENT:
+ call_conv = RADEON_LLVM_AMDGPU_PS;
+ break;
+ case PIPE_SHADER_COMPUTE:
+ call_conv = RADEON_LLVM_AMDGPU_CS;
+ break;
+ default:
+ unreachable("Unhandle shader type");
+ }
+
+ LLVMSetFunctionCallConv(ctx->main_fn, call_conv);
}
void si_llvm_optimize_module(struct si_shader_context *ctx)
{
struct gallivm_state *gallivm = &ctx->gallivm;
const char *triple = LLVMGetTarget(gallivm->module);
LLVMTargetLibraryInfoRef target_library_info;
/* Dump LLVM IR before any optimization passes */
if (ctx->screen->b.debug_flags & DBG_PREOPT_IR &&
--
2.7.4
_______________________________________________
mesa-dev mailing list
mesa-dev@lists.freedesktop.org
https://lists.freedesktop.org/mailman/listinfo/mesa-dev
