summaryrefslogtreecommitdiffstats
path: root/clang
diff options
context:
space:
mode:
Diffstat (limited to 'clang')
-rw-r--r--clang/lib/CodeGen/CGOpenMPRuntime.cpp47
-rw-r--r--clang/test/OpenMP/target_codegen.cpp7
-rw-r--r--clang/test/OpenMP/target_messages.cpp21
3 files changed, 63 insertions, 12 deletions
diff --git a/clang/lib/CodeGen/CGOpenMPRuntime.cpp b/clang/lib/CodeGen/CGOpenMPRuntime.cpp
index d301a4c1675..33dfe71dff0 100644
--- a/clang/lib/CodeGen/CGOpenMPRuntime.cpp
+++ b/clang/lib/CodeGen/CGOpenMPRuntime.cpp
@@ -2588,21 +2588,20 @@ llvm::Function *CGOpenMPRuntime::emitThreadPrivateVarDefinition(
static void getTargetEntryUniqueInfo(ASTContext &C, SourceLocation Loc,
unsigned &DeviceID, unsigned &FileID,
unsigned &LineNum) {
-
SourceManager &SM = C.getSourceManager();
// The loc should be always valid and have a file ID (the user cannot use
// #pragma directives in macros)
assert(Loc.isValid() && "Source location is expected to be always valid.");
- assert(Loc.isFileID() && "Source location is expected to refer to a file.");
PresumedLoc PLoc = SM.getPresumedLoc(Loc);
assert(PLoc.isValid() && "Source location is expected to be always valid.");
llvm::sys::fs::UniqueID ID;
- if (llvm::sys::fs::getUniqueID(PLoc.getFilename(), ID))
- llvm_unreachable("Source file with target region no longer exists!");
+ if (auto EC = llvm::sys::fs::getUniqueID(PLoc.getFilename(), ID))
+ SM.getDiagnostics().Report(diag::err_cannot_open_file)
+ << PLoc.getFilename() << EC.message();
DeviceID = ID.getDevice();
FileID = ID.getFile();
@@ -3586,8 +3585,13 @@ void CGOpenMPRuntime::OffloadEntriesInfoManagerTy::
// If we are emitting code for a target, the entry is already initialized,
// only has to be registered.
if (CGM.getLangOpts().OpenMPIsDevice) {
- assert(hasTargetRegionEntryInfo(DeviceID, FileID, ParentName, LineNum) &&
- "Entry must exist.");
+ if (!hasTargetRegionEntryInfo(DeviceID, FileID, ParentName, LineNum)) {
+ unsigned DiagID = CGM.getDiags().getCustomDiagID(
+ DiagnosticsEngine::Error,
+ "Unable to find target region on line '%0' in the device code.");
+ CGM.getDiags().Report(DiagID) << LineNum;
+ return;
+ }
auto &Entry =
OffloadEntriesTargetRegion[DeviceID][FileID][ParentName][LineNum];
assert(Entry.isValid() && "Entry not initialized!");
@@ -3928,14 +3932,27 @@ void CGOpenMPRuntime::createOffloadEntriesAndInfoMetadata() {
if (const auto *CE =
dyn_cast<OffloadEntriesInfoManagerTy::OffloadEntryInfoTargetRegion>(
E)) {
- assert(CE->getID() && CE->getAddress() &&
- "Entry ID and Addr are invalid!");
+ if (!CE->getID() || !CE->getAddress()) {
+ unsigned DiagID = CGM.getDiags().getCustomDiagID(
+ DiagnosticsEngine::Error,
+ "Offloading entry for target region is incorect: either the "
+ "address or the ID is invalid.");
+ CGM.getDiags().Report(DiagID);
+ continue;
+ }
createOffloadEntry(CE->getID(), CE->getAddress(), /*Size=*/0,
CE->getFlags(), llvm::GlobalValue::WeakAnyLinkage);
} else if (const auto *CE =
dyn_cast<OffloadEntriesInfoManagerTy::
OffloadEntryInfoDeviceGlobalVar>(E)) {
- assert(CE->getAddress() && "Entry Addr is invalid!");
+ if (!CE->getAddress()) {
+ unsigned DiagID = CGM.getDiags().getCustomDiagID(
+ DiagnosticsEngine::Error,
+ "Offloading entry for declare target varible is inccorect: the "
+ "address is invalid.");
+ CGM.getDiags().Report(DiagID);
+ continue;
+ }
createOffloadEntry(CE->getAddress(), CE->getAddress(),
CE->getVarSize().getQuantity(), CE->getFlags(),
CE->getLinkage());
@@ -3958,15 +3975,23 @@ void CGOpenMPRuntime::loadOffloadInfoMetadata() {
return;
auto Buf = llvm::MemoryBuffer::getFile(CGM.getLangOpts().OMPHostIRFile);
- if (Buf.getError())
+ if (auto EC = Buf.getError()) {
+ CGM.getDiags().Report(diag::err_cannot_open_file)
+ << CGM.getLangOpts().OMPHostIRFile << EC.message();
return;
+ }
llvm::LLVMContext C;
auto ME = expectedToErrorOrAndEmitErrors(
C, llvm::parseBitcodeFile(Buf.get()->getMemBufferRef(), C));
- if (ME.getError())
+ if (auto EC = ME.getError()) {
+ unsigned DiagID = CGM.getDiags().getCustomDiagID(
+ DiagnosticsEngine::Error, "Unable to parse host IR file '%0':'%1'");
+ CGM.getDiags().Report(DiagID)
+ << CGM.getLangOpts().OMPHostIRFile << EC.message();
return;
+ }
llvm::NamedMDNode *MD = ME.get()->getNamedMetadata("omp_offload.info");
if (!MD)
diff --git a/clang/test/OpenMP/target_codegen.cpp b/clang/test/OpenMP/target_codegen.cpp
index e75ea738c8e..618a5be995b 100644
--- a/clang/test/OpenMP/target_codegen.cpp
+++ b/clang/test/OpenMP/target_codegen.cpp
@@ -80,6 +80,7 @@
// TCHECK: @{{.+}} = weak constant [[ENTTY]]
// TCHECK: @{{.+}} = weak constant [[ENTTY]]
// TCHECK: @{{.+}} = {{.*}}constant [[ENTTY]]
+// TCHECK: @{{.+}} = {{.*}}constant [[ENTTY]]
// TCHECK-NOT: @{{.+}} = weak constant [[ENTTY]]
// Check if offloading descriptor is created.
@@ -750,4 +751,10 @@ int bar(int n){
// CHECK-32-DAG: load i32, i32* [[LOCAL_A]]
// CHECK-DAG: load i16, i16* [[REF_AA]]
// CHECK-DAG: getelementptr inbounds [10 x i32], [10 x i32]* [[REF_B]], i[[SZ]] 0, i[[SZ]] 2
+
+void bar () {
+#define pragma_target _Pragma("omp target")
+pragma_target
+{}
+}
#endif
diff --git a/clang/test/OpenMP/target_messages.cpp b/clang/test/OpenMP/target_messages.cpp
index 25db02690c3..52dbe4900d6 100644
--- a/clang/test/OpenMP/target_messages.cpp
+++ b/clang/test/OpenMP/target_messages.cpp
@@ -9,6 +9,25 @@
// RUN: not %clang_cc1 -fopenmp -std=c++11 -fopenmp-targets=hexagon-linux-gnu -o - %s 2>&1 | FileCheck --check-prefix CHECK-UNSUPPORTED-DEVICE-TARGET %s
// CHECK-UNSUPPORTED-DEVICE-TARGET: OpenMP target is invalid: 'hexagon-linux-gnu'
+// RUN: not %clang_cc1 -fopenmp -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path 1111.bc -o - 2>&1 | FileCheck --check-prefix NO-HOST-BC %s
+// NO-HOST-BC: The provided host compiler IR file '1111.bc' is required to generate code for OpenMP target regions but cannot be found.
+
+// RUN: %clang_cc1 -fopenmp -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm-bc %s -o %t-ppc-host.bc -DREGION_HOST
+// RUN: not %clang_cc1 -fopenmp -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o - -DREGION_DEVICE 2>&1 | FileCheck %s --check-prefix NO-REGION
+// NO-REGION: Offloading entry for target region is incorect: either the address or the ID is invalid.
+
+#if defined(REGION_HOST) || defined(REGION_DEVICE)
+void foo() {
+#ifdef REGION_HOST
+#pragma omp target
+ ;
+#endif
+#ifdef REGION_DEVICE
+#pragma omp target
+ ;
+#endif
+}
+#else
void foo() {
}
@@ -71,4 +90,4 @@ int main(int argc, char **argv) {
return 0;
}
-
+#endif
OpenPOWER on IntegriCloud