Author: abataev
Date: Mon Apr 30 09:26:57 2018
New Revision: 331195
URL: http://llvm.org/viewvc/llvm-project?rev=331195&view=rev
Log:
[OPENMP] Do not crash on incorrect input data.
Emit error messages instead of compiler crashing when the target region
does not exist in the device code + fix crash when the location comes
from macros.
Modified:
cfe/trunk/lib/CodeGen/CGOpenMPRuntime.cpp
cfe/trunk/test/OpenMP/target_codegen.cpp
cfe/trunk/test/OpenMP/target_messages.cpp
Modified: cfe/trunk/lib/CodeGen/CGOpenMPRuntime.cpp
URL:
http://llvm.org/viewvc/llvm-project/cfe/trunk/lib/CodeGen/CGOpenMPRuntime.cpp?rev=331195&r1=331194&r2=331195&view=diff
==============================================================================
--- cfe/trunk/lib/CodeGen/CGOpenMPRuntime.cpp (original)
+++ cfe/trunk/lib/CodeGen/CGOpenMPRuntime.cpp Mon Apr 30 09:26:57 2018
@@ -2588,21 +2588,20 @@ llvm::Function *CGOpenMPRuntime::emitThr
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::OffloadEntriesInfo
// 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::createOffloadEntri
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::loadOffloadInfoMet
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)
Modified: cfe/trunk/test/OpenMP/target_codegen.cpp
URL:
http://llvm.org/viewvc/llvm-project/cfe/trunk/test/OpenMP/target_codegen.cpp?rev=331195&r1=331194&r2=331195&view=diff
==============================================================================
--- cfe/trunk/test/OpenMP/target_codegen.cpp (original)
+++ cfe/trunk/test/OpenMP/target_codegen.cpp Mon Apr 30 09:26:57 2018
@@ -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
Modified: cfe/trunk/test/OpenMP/target_messages.cpp
URL:
http://llvm.org/viewvc/llvm-project/cfe/trunk/test/OpenMP/target_messages.cpp?rev=331195&r1=331194&r2=331195&view=diff
==============================================================================
--- cfe/trunk/test/OpenMP/target_messages.cpp (original)
+++ cfe/trunk/test/OpenMP/target_messages.cpp Mon Apr 30 09:26:57 2018
@@ -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
_______________________________________________
cfe-commits mailing list
cfe-commits@lists.llvm.org
http://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
