ac/llvm: inline ac_get_load_intr_attribs

Reviewed-by: Pierre-Eric Pelloux-Prayer <pierre-eric.pelloux-prayer@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/20146>

diff --git a/src/amd/llvm/ac_llvm_build.c b/src/amd/llvm/ac_llvm_build.c
index bf7c25c..d46b732 100644 (file)
--- a/src/amd/llvm/ac_llvm_build.c
+++ b/src/amd/llvm/ac_llvm_build.c
@@ -1322,7 +1322,8 @@ static LLVMValueRef ac_build_buffer_load_common(struct ac_llvm_context *ctx, LLV
       snprintf(name, sizeof(name), "llvm.amdgcn.%s.buffer.load.%s", indexing_kind, type_name);
    }
 
-   return ac_build_intrinsic(ctx, name, type, args, idx, ac_get_load_intr_attribs(can_speculate));
+   return ac_build_intrinsic(ctx, name, type, args, idx,
+                             can_speculate ? AC_ATTR_INVARIANT_LOAD : 0);
 }
 
 LLVMValueRef ac_build_buffer_load(struct ac_llvm_context *ctx, LLVMValueRef rsrc, int num_channels,
@@ -1435,7 +1436,8 @@ static LLVMValueRef ac_build_tbuffer_load(struct ac_llvm_context *ctx, LLVMValue
 
    snprintf(name, sizeof(name), "llvm.amdgcn.%s.tbuffer.load.%s", indexing_kind, type_name);
 
-   return ac_build_intrinsic(ctx, name, type, args, idx, ac_get_load_intr_attribs(can_speculate));
+   return ac_build_intrinsic(ctx, name, type, args, idx,
+                             can_speculate ? AC_ATTR_INVARIANT_LOAD : 0);
 }
 
 LLVMValueRef ac_build_struct_tbuffer_load(struct ac_llvm_context *ctx, LLVMValueRef rsrc,

diff --git a/src/amd/llvm/ac_llvm_util.h b/src/amd/llvm/ac_llvm_util.h
index c8bd6e7..5006950 100644 (file)
--- a/src/amd/llvm/ac_llvm_util.h
+++ b/src/amd/llvm/ac_llvm_util.h
@@ -96,11 +96,6 @@ void ac_llvm_add_target_dep_function_attr(LLVMValueRef F, const char *name, unsi
 void ac_llvm_set_workgroup_size(LLVMValueRef F, unsigned size);
 void ac_llvm_set_target_features(LLVMValueRef F, struct ac_llvm_context *ctx);
 
-static inline unsigned ac_get_load_intr_attribs(bool can_speculate)
-{
-   return can_speculate ? AC_ATTR_INVARIANT_LOAD : 0;
-}
-
 LLVMTargetLibraryInfoRef ac_create_target_library_info(const char *triple);
 void ac_dispose_target_library_info(LLVMTargetLibraryInfoRef library_info);
 PUBLIC void ac_init_shared_llvm_once(void); /* Do not use directly, use ac_init_llvm_once */