Mesa (master): radeonsi: fix passing address32_hi to LLVM for high values
Marek Olšák
mareko at kemper.freedesktop.org
Wed Mar 7 18:56:06 UTC 2018
Module: Mesa
Branch: master
Commit: 2c3f3651c4b871e4c9116fe3d8c2e447ae1d9141
URL: http://cgit.freedesktop.org/mesa/mesa/commit/?id=2c3f3651c4b871e4c9116fe3d8c2e447ae1d9141
Author: Marek Olšák <marek.olsak at amd.com>
Date: Sun Feb 25 00:33:28 2018 +0100
radeonsi: fix passing address32_hi to LLVM for high values
The old function treats high values as negative, which LLVM interprets as 0.
---
src/amd/common/ac_llvm_util.c | 4 ++--
src/amd/common/ac_llvm_util.h | 2 +-
src/gallium/drivers/radeonsi/si_shader.c | 7 +++++--
3 files changed, 8 insertions(+), 5 deletions(-)
diff --git a/src/amd/common/ac_llvm_util.c b/src/amd/common/ac_llvm_util.c
index 3530bf088b..bb9e873af8 100644
--- a/src/amd/common/ac_llvm_util.c
+++ b/src/amd/common/ac_llvm_util.c
@@ -202,11 +202,11 @@ ac_dump_module(LLVMModuleRef module)
void
ac_llvm_add_target_dep_function_attr(LLVMValueRef F,
- const char *name, int value)
+ const char *name, unsigned value)
{
char str[16];
- snprintf(str, sizeof(str), "%i", value);
+ snprintf(str, sizeof(str), "0x%x", value);
LLVMAddTargetDependentFunctionAttr(F, name, str);
}
diff --git a/src/amd/common/ac_llvm_util.h b/src/amd/common/ac_llvm_util.h
index 5329bb1b70..9c6b89bf6c 100644
--- a/src/amd/common/ac_llvm_util.h
+++ b/src/amd/common/ac_llvm_util.h
@@ -87,7 +87,7 @@ LLVMBuilderRef ac_create_builder(LLVMContextRef ctx,
void
ac_llvm_add_target_dep_function_attr(LLVMValueRef F,
- const char *name, int value);
+ const char *name, unsigned value);
static inline unsigned
ac_get_load_intr_attribs(bool can_speculate)
diff --git a/src/gallium/drivers/radeonsi/si_shader.c b/src/gallium/drivers/radeonsi/si_shader.c
index 2e57eca6e5..d95e69f81a 100644
--- a/src/gallium/drivers/radeonsi/si_shader.c
+++ b/src/gallium/drivers/radeonsi/si_shader.c
@@ -4456,8 +4456,11 @@ static void si_create_function(struct si_shader_context *ctx,
*fninfo->assign[i] = LLVMGetParam(ctx->main_fn, i);
}
- si_llvm_add_attribute(ctx->main_fn, "amdgpu-32bit-address-high-bits",
- ctx->screen->info.address32_hi);
+ if (ctx->screen->info.address32_hi) {
+ ac_llvm_add_target_dep_function_attr(ctx->main_fn,
+ "amdgpu-32bit-address-high-bits",
+ ctx->screen->info.address32_hi);
+ }
if (max_workgroup_size) {
si_llvm_add_attribute(ctx->main_fn, "amdgpu-max-work-group-size",
More information about the mesa-commit
mailing list