[PATCH] D110556: [openmp] Add addrspacecast to getOrCreateIdent

Jon Chesterfield via Phabricator via llvm-commits llvm-commits at lists.llvm.org
Mon Sep 27 10:55:34 PDT 2021


JonChesterfield updated this revision to Diff 375340.
JonChesterfield added a comment.

- Set addrspace on the nearby variable, revert to CreatePointerCast


Repository:
  rG LLVM Github Monorepo

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

https://reviews.llvm.org/D110556

Files:
  llvm/lib/Frontend/OpenMP/OMPIRBuilder.cpp
  openmp/libomptarget/test/offloading/bug51982.c


Index: openmp/libomptarget/test/offloading/bug51982.c
===================================================================
--- /dev/null
+++ openmp/libomptarget/test/offloading/bug51982.c
@@ -0,0 +1,24 @@
+// RUN: %libomptarget-compile-generic -O1 && %libomptarget-run-generic
+// -O1 to run openmp-opt
+
+int main(void) {
+  long int aa = 0;
+
+  int ng = 12;
+  int nxyz = 5;
+
+  const long exp = ng * nxyz;
+
+#pragma omp target map(tofrom : aa)
+  for (int gid = 0; gid < nxyz; gid++) {
+#pragma omp parallel for
+    for (unsigned int g = 0; g < ng; g++) {
+#pragma omp atomic
+      aa += 1;
+    }
+  }
+  if (aa != exp) {
+    return 1;
+  }
+  return 0;
+}
Index: llvm/lib/Frontend/OpenMP/OMPIRBuilder.cpp
===================================================================
--- llvm/lib/Frontend/OpenMP/OMPIRBuilder.cpp
+++ llvm/lib/Frontend/OpenMP/OMPIRBuilder.cpp
@@ -276,15 +276,20 @@
     for (GlobalVariable &GV : M.getGlobalList())
       if (GV.getValueType() == OpenMPIRBuilder::Ident && GV.hasInitializer())
         if (GV.getInitializer() == Initializer)
-          return Ident = &GV;
-
-    auto *GV = new GlobalVariable(M, OpenMPIRBuilder::Ident,
-                                  /* isConstant = */ true,
-                                  GlobalValue::PrivateLinkage, Initializer);
-    GV->setUnnamedAddr(GlobalValue::UnnamedAddr::Global);
-    GV->setAlignment(Align(8));
-    Ident = GV;
+          Ident = &GV;
+
+    if (!Ident) {
+      auto *GV = new GlobalVariable(
+          M, OpenMPIRBuilder::Ident,
+          /* isConstant = */ true, GlobalValue::PrivateLinkage, Initializer, "",
+          nullptr, GlobalValue::NotThreadLocal,
+          M.getDataLayout().getDefaultGlobalsAddressSpace());
+      GV->setUnnamedAddr(GlobalValue::UnnamedAddr::Global);
+      GV->setAlignment(Align(8));
+      Ident = GV;
+    }
   }
+
   return Builder.CreatePointerCast(Ident, IdentPtr);
 }
 


-------------- next part --------------
A non-text attachment was scrubbed...
Name: D110556.375340.patch
Type: text/x-patch
Size: 1920 bytes
Desc: not available
URL: <http://lists.llvm.org/pipermail/llvm-commits/attachments/20210927/8485deb9/attachment.bin>


More information about the llvm-commits mailing list