From 80fa43fe9ab41e4e035547ab0f1955efa81de3a2 Mon Sep 17 00:00:00 2001 From: Jon Chesterfield Date: Mon, 27 Sep 2021 19:27:00 +0100 Subject: [PATCH] Revert "[openmp] Add addrspacecast to getOrCreateIdent" This reverts commit 1a761e5b7b50dc08e0ff7f7aea65e1da29c5cd80. Failed CI, albeit with a different failure mode to BZ51982 --- llvm/lib/Frontend/OpenMP/OMPIRBuilder.cpp | 19 ++++++--------- .../libomptarget/test/offloading/bug51982.c | 24 ------------------- 2 files changed, 7 insertions(+), 36 deletions(-) delete mode 100644 openmp/libomptarget/test/offloading/bug51982.c diff --git a/llvm/lib/Frontend/OpenMP/OMPIRBuilder.cpp b/llvm/lib/Frontend/OpenMP/OMPIRBuilder.cpp index 36f6fa68ce24..94217c632e43 100644 --- a/llvm/lib/Frontend/OpenMP/OMPIRBuilder.cpp +++ b/llvm/lib/Frontend/OpenMP/OMPIRBuilder.cpp @@ -276,20 +276,15 @@ Value *OpenMPIRBuilder::getOrCreateIdent(Constant *SrcLocStr, for (GlobalVariable &GV : M.getGlobalList()) if (GV.getValueType() == OpenMPIRBuilder::Ident && GV.hasInitializer()) if (GV.getInitializer() == Initializer) - Ident = &GV; + return 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; - } + auto *GV = new GlobalVariable(M, OpenMPIRBuilder::Ident, + /* isConstant = */ true, + GlobalValue::PrivateLinkage, Initializer); + GV->setUnnamedAddr(GlobalValue::UnnamedAddr::Global); + GV->setAlignment(Align(8)); + Ident = GV; } - return Builder.CreatePointerCast(Ident, IdentPtr); } diff --git a/openmp/libomptarget/test/offloading/bug51982.c b/openmp/libomptarget/test/offloading/bug51982.c deleted file mode 100644 index 4211190ec2dd..000000000000 --- a/openmp/libomptarget/test/offloading/bug51982.c +++ /dev/null @@ -1,24 +0,0 @@ -// 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; -}