[PATCH] D117320: [OpenMP] Mark device RTL variables as hidden

[Joseph Huber via Phabricator via llvm-commits]

jhuber6 updated this revision to Diff 400137.
jhuber6 added a comment.
Herald added a project: clang.
Herald added a subscriber: cfe-commits.

Fix test and add the fact that its hidden to the comment.


Repository:
  rG LLVM Github Monorepo

CHANGES SINCE LAST ACTION
  https://reviews.llvm.org/D117320/new/

https://reviews.llvm.org/D117320

Files:
  clang/test/OpenMP/target_globals_codegen.cpp
  llvm/include/llvm/Frontend/OpenMP/OMPIRBuilder.h
  llvm/lib/Frontend/OpenMP/OMPIRBuilder.cpp


Index: llvm/lib/Frontend/OpenMP/OMPIRBuilder.cpp
===================================================================
--- llvm/lib/Frontend/OpenMP/OMPIRBuilder.cpp
+++ llvm/lib/Frontend/OpenMP/OMPIRBuilder.cpp
@@ -254,6 +254,7 @@
       new GlobalVariable(M, I32Ty,
                          /* isConstant = */ true, GlobalValue::WeakODRLinkage,
                          ConstantInt::get(I32Ty, Value), Name);
+  GV->setVisibility(GlobalValue::HiddenVisibility);
 
   return GV;
 }
Index: llvm/include/llvm/Frontend/OpenMP/OMPIRBuilder.h
===================================================================
--- llvm/include/llvm/Frontend/OpenMP/OMPIRBuilder.h
+++ llvm/include/llvm/Frontend/OpenMP/OMPIRBuilder.h
@@ -689,7 +689,8 @@
                              omp::IdentFlag Flags = omp::IdentFlag(0),
                              unsigned Reserve2Flags = 0);
 
-  /// Create a global flag \p Namein the module with initial value \p Value.
+  /// Create a hidden global flag \p Name in the module with initial value \p
+  /// Value.
   GlobalValue *createGlobalFlag(unsigned Value, StringRef Name);
 
   /// Generate control flow and cleanup for cancellation.
Index: clang/test/OpenMP/target_globals_codegen.cpp
===================================================================
--- clang/test/OpenMP/target_globals_codegen.cpp
+++ clang/test/OpenMP/target_globals_codegen.cpp
@@ -12,25 +12,25 @@
 #define HEADER
 
 //.
-// CHECK: @__omp_rtl_debug_kind = weak_odr constant i32 1
-// CHECK: @__omp_rtl_assume_teams_oversubscription = weak_odr constant i32 0
-// CHECK: @__omp_rtl_assume_threads_oversubscription = weak_odr constant i32 0
+// CHECK: @__omp_rtl_debug_kind = weak_odr hidden constant i32 1
+// CHECK: @__omp_rtl_assume_teams_oversubscription = weak_odr hidden constant i32 0
+// CHECK: @__omp_rtl_assume_threads_oversubscription = weak_odr hidden constant i32 0
 //.
-// CHECK-EQ: @__omp_rtl_debug_kind = weak_odr constant i32 111
-// CHECK-EQ: @__omp_rtl_assume_teams_oversubscription = weak_odr constant i32 0
-// CHECK-EQ: @__omp_rtl_assume_threads_oversubscription = weak_odr constant i32 0
+// CHECK-EQ: @__omp_rtl_debug_kind = weak_odr hidden constant i32 111
+// CHECK-EQ: @__omp_rtl_assume_teams_oversubscription = weak_odr hidden constant i32 0
+// CHECK-EQ: @__omp_rtl_assume_threads_oversubscription = weak_odr hidden constant i32 0
 //.
-// CHECK-DEFAULT: @__omp_rtl_debug_kind = weak_odr constant i32 0
-// CHECK-DEFAULT: @__omp_rtl_assume_teams_oversubscription = weak_odr constant i32 0
-// CHECK-DEFAULT: @__omp_rtl_assume_threads_oversubscription = weak_odr constant i32 0
+// CHECK-DEFAULT: @__omp_rtl_debug_kind = weak_odr hidden constant i32 0
+// CHECK-DEFAULT: @__omp_rtl_assume_teams_oversubscription = weak_odr hidden constant i32 0
+// CHECK-DEFAULT: @__omp_rtl_assume_threads_oversubscription = weak_odr hidden constant i32 0
 //.
-// CHECK-THREADS: @__omp_rtl_debug_kind = weak_odr constant i32 0
-// CHECK-THREADS: @__omp_rtl_assume_teams_oversubscription = weak_odr constant i32 0
-// CHECK-THREADS: @__omp_rtl_assume_threads_oversubscription = weak_odr constant i32 1
+// CHECK-THREADS: @__omp_rtl_debug_kind = weak_odr hidden constant i32 0
+// CHECK-THREADS: @__omp_rtl_assume_teams_oversubscription = weak_odr hidden constant i32 0
+// CHECK-THREADS: @__omp_rtl_assume_threads_oversubscription = weak_odr hidden constant i32 1
 //.
-// CHECK-TEAMS: @__omp_rtl_debug_kind = weak_odr constant i32 0
-// CHECK-TEAMS: @__omp_rtl_assume_teams_oversubscription = weak_odr constant i32 1
-// CHECK-TEAMS: @__omp_rtl_assume_threads_oversubscription = weak_odr constant i32 0
+// CHECK-TEAMS: @__omp_rtl_debug_kind = weak_odr hidden constant i32 0
+// CHECK-TEAMS: @__omp_rtl_assume_teams_oversubscription = weak_odr hidden constant i32 1
+// CHECK-TEAMS: @__omp_rtl_assume_threads_oversubscription = weak_odr hidden constant i32 0
 //.
 void foo() {
 #pragma omp target


-------------- next part --------------
A non-text attachment was scrubbed...
Name: D117320.400137.patch
Type: text/x-patch
Size: 3912 bytes
Desc: not available
URL: <http://lists.llvm.org/pipermail/llvm-commits/attachments/20220114/3a589254/attachment.bin>