diff options
| author | Marek Olšák <[email protected]> | 2018-02-25 00:33:28 +0100 |
|---|---|---|
| committer | Marek Olšák <[email protected]> | 2018-03-07 13:55:49 -0500 |
| commit | 2c3f3651c4b871e4c9116fe3d8c2e447ae1d9141 (patch) | |
| tree | 04346fff29a4c8527197e5bd8529dcf15f07f4a0 /src | |
| parent | b3b6b00ac8b7cda04ebbad71f256a57071714cf5 (diff) | |
radeonsi: fix passing address32_hi to LLVM for high values
The old function treats high values as negative, which LLVM interprets as 0.
Diffstat (limited to 'src')
| -rw-r--r-- | src/amd/common/ac_llvm_util.c | 4 | ||||
| -rw-r--r-- | src/amd/common/ac_llvm_util.h | 2 | ||||
| -rw-r--r-- | 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 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", |