From: Marek Olšák <marek.olsak@amd.com>
Date: Sat, 24 Feb 2018 23:33:28 +0000 (+0100)
Subject: radeonsi: fix passing address32_hi to LLVM for high values
X-Git-Url: https://git.libre-soc.org/?a=commitdiff_plain;h=2c3f3651c4b871e4c9116fe3d8c2e447ae1d9141;p=mesa.git

radeonsi: fix passing address32_hi to LLVM for high values

The old function treats high values as negative, which LLVM interprets as 0.
---

diff --git a/src/amd/common/ac_llvm_util.c b/src/amd/common/ac_llvm_util.c
index 3530bf088be..bb9e873af81 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 5329bb1b702..9c6b89bf6c1 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 2e57eca6e54..d95e69f81ad 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",