summaryrefslogtreecommitdiffstats
path: root/src/gallium/drivers
diff options
context:
space:
mode:
authorMarek Olšák <[email protected]>2018-08-13 19:11:55 -0400
committerMarek Olšák <[email protected]>2019-05-16 13:06:40 -0400
commitc7ceeea0937f1844d080fce347aec393b8a13a85 (patch)
treeb362630fb8152fc20ab6bc5e55bb939902d3c0ec /src/gallium/drivers
parentd569b7cb31160e1cacdda3663f0d4249245de00d (diff)
radeonsi: return the last part's return value from @wrapper
The primitive discard compute shader will get the position output this way. Tested-by: Dieter Nützel <[email protected]> Acked-by: Nicolai Hähnle <[email protected]>
Diffstat (limited to 'src/gallium/drivers')
-rw-r--r--src/gallium/drivers/radeonsi/si_shader.c29
1 files changed, 26 insertions, 3 deletions
diff --git a/src/gallium/drivers/radeonsi/si_shader.c b/src/gallium/drivers/radeonsi/si_shader.c
index 98c11e1c98d..6455bb5dcd0 100644
--- a/src/gallium/drivers/radeonsi/si_shader.c
+++ b/src/gallium/drivers/radeonsi/si_shader.c
@@ -6399,7 +6399,26 @@ static void si_build_wrapper_function(struct si_shader_context *ctx,
gprs += size;
}
- si_create_function(ctx, "wrapper", NULL, 0, &fninfo,
+ /* Prepare the return type. */
+ unsigned num_returns = 0;
+ LLVMTypeRef returns[32], last_func_type, return_type;
+
+ last_func_type = LLVMGetElementType(LLVMTypeOf(parts[num_parts - 1]));
+ return_type = LLVMGetReturnType(last_func_type);
+
+ switch (LLVMGetTypeKind(return_type)) {
+ case LLVMStructTypeKind:
+ num_returns = LLVMCountStructElementTypes(return_type);
+ assert(num_returns <= ARRAY_SIZE(returns));
+ LLVMGetStructElementTypes(return_type, returns);
+ break;
+ case LLVMVoidTypeKind:
+ break;
+ default:
+ unreachable("unexpected type");
+ }
+
+ si_create_function(ctx, "wrapper", returns, num_returns, &fninfo,
si_get_max_workgroup_size(ctx->shader));
if (is_merged_shader(ctx))
@@ -6451,9 +6470,9 @@ static void si_build_wrapper_function(struct si_shader_context *ctx,
initial_num_out_sgpr = num_out_sgpr;
/* Now chain the parts. */
+ LLVMValueRef ret;
for (unsigned part = 0; part < num_parts; ++part) {
LLVMValueRef in[48];
- LLVMValueRef ret;
LLVMTypeRef ret_type;
unsigned out_idx = 0;
unsigned num_params = LLVMCountParams(parts[part]);
@@ -6565,7 +6584,11 @@ static void si_build_wrapper_function(struct si_shader_context *ctx,
}
}
- LLVMBuildRetVoid(builder);
+ /* Return the value from the last part. */
+ if (LLVMGetTypeKind(LLVMTypeOf(ret)) == LLVMVoidTypeKind)
+ LLVMBuildRetVoid(builder);
+ else
+ LLVMBuildRet(builder, ret);
}
static bool si_should_optimize_less(struct ac_llvm_compiler *compiler,