From: Alexey Bataev Date: Thu, 3 Oct 2019 16:20:34 +0000 (+0000) Subject: [OPENMP]Improve diagnostics for not found declare target entries. X-Git-Url: https://granicus.if.org/sourcecode?a=commitdiff_plain;h=a890b574da851b2367279d7aebdb28ce478e2b35;p=clang [OPENMP]Improve diagnostics for not found declare target entries. We can point to the target region + emit parent functions names/real var names if they were not found in host module during device codegen. git-svn-id: https://llvm.org/svn/llvm-project/cfe/trunk@373620 91177308-0d34-0410-b5e6-96231b3b80d8 --- diff --git a/lib/CodeGen/CGOpenMPRuntime.cpp b/lib/CodeGen/CGOpenMPRuntime.cpp index 4da84f230e..f64f39defb 100644 --- a/lib/CodeGen/CGOpenMPRuntime.cpp +++ b/lib/CodeGen/CGOpenMPRuntime.cpp @@ -4201,7 +4201,9 @@ void CGOpenMPRuntime::createOffloadEntriesAndInfoMetadata() { llvm::Module &M = CGM.getModule(); llvm::LLVMContext &C = M.getContext(); - SmallVector + SmallVector, + 16> OrderedEntries(OffloadEntriesInfoManager.size()); llvm::SmallVector ParentFunctions( OffloadEntriesInfoManager.size()); @@ -4219,7 +4221,8 @@ void CGOpenMPRuntime::createOffloadEntriesAndInfoMetadata() { // Create function that emits metadata for each target region entry; auto &&TargetRegionMetadataEmitter = - [&C, MD, &OrderedEntries, &ParentFunctions, &GetMDInt, &GetMDString]( + [this, &C, MD, &OrderedEntries, &ParentFunctions, &GetMDInt, + &GetMDString]( unsigned DeviceID, unsigned FileID, StringRef ParentName, unsigned Line, const OffloadEntriesInfoManagerTy::OffloadEntryInfoTargetRegion &E) { @@ -4237,8 +4240,19 @@ void CGOpenMPRuntime::createOffloadEntriesAndInfoMetadata() { GetMDInt(FileID), GetMDString(ParentName), GetMDInt(Line), GetMDInt(E.getOrder())}; + SourceLocation Loc; + for (auto I = CGM.getContext().getSourceManager().fileinfo_begin(), + E = CGM.getContext().getSourceManager().fileinfo_end(); + I != E; ++I) { + if (I->getFirst()->getUniqueID().getDevice() == DeviceID && + I->getFirst()->getUniqueID().getFile() == FileID) { + Loc = CGM.getContext().getSourceManager().translateFileLineCol( + I->getFirst(), Line, 1); + break; + } + } // Save this entry in the right position of the ordered entries array. - OrderedEntries[E.getOrder()] = &E; + OrderedEntries[E.getOrder()] = std::make_tuple(&E, Loc, ParentName); ParentFunctions[E.getOrder()] = ParentName; // Add metadata to the named metadata node. @@ -4266,7 +4280,8 @@ void CGOpenMPRuntime::createOffloadEntriesAndInfoMetadata() { GetMDInt(E.getFlags()), GetMDInt(E.getOrder())}; // Save this entry in the right position of the ordered entries array. - OrderedEntries[E.getOrder()] = &E; + OrderedEntries[E.getOrder()] = + std::make_tuple(&E, SourceLocation(), MangledName); // Add metadata to the named metadata node. MD->addOperand(llvm::MDNode::get(C, Ops)); @@ -4275,11 +4290,11 @@ void CGOpenMPRuntime::createOffloadEntriesAndInfoMetadata() { OffloadEntriesInfoManager.actOnDeviceGlobalVarEntriesInfo( DeviceGlobalVarMetadataEmitter); - for (const auto *E : OrderedEntries) { - assert(E && "All ordered entries must exist!"); + for (const auto &E : OrderedEntries) { + assert(std::get<0>(E) && "All ordered entries must exist!"); if (const auto *CE = dyn_cast( - E)) { + std::get<0>(E))) { if (!CE->getID() || !CE->getAddress()) { // Do not blame the entry if the parent funtion is not emitted. StringRef FnName = ParentFunctions[CE->getOrder()]; @@ -4287,16 +4302,16 @@ void CGOpenMPRuntime::createOffloadEntriesAndInfoMetadata() { continue; unsigned DiagID = CGM.getDiags().getCustomDiagID( DiagnosticsEngine::Error, - "Offloading entry for target region is incorrect: either the " + "Offloading entry for target region in %0 is incorrect: either the " "address or the ID is invalid."); - CGM.getDiags().Report(DiagID); + CGM.getDiags().Report(std::get<1>(E), DiagID) << FnName; continue; } createOffloadEntry(CE->getID(), CE->getAddress(), /*Size=*/0, CE->getFlags(), llvm::GlobalValue::WeakAnyLinkage); - } else if (const auto *CE = - dyn_cast(E)) { + } else if (const auto *CE = dyn_cast( + std::get<0>(E))) { OffloadEntriesInfoManagerTy::OMPTargetGlobalVarEntryKind Flags = static_cast( CE->getFlags()); @@ -4307,10 +4322,10 @@ void CGOpenMPRuntime::createOffloadEntriesAndInfoMetadata() { continue; if (!CE->getAddress()) { unsigned DiagID = CGM.getDiags().getCustomDiagID( - DiagnosticsEngine::Error, - "Offloading entry for declare target variable is incorrect: the " - "address is invalid."); - CGM.getDiags().Report(DiagID); + DiagnosticsEngine::Error, "Offloading entry for declare target " + "variable %0 is incorrect: the " + "address is invalid."); + CGM.getDiags().Report(std::get<1>(E), DiagID) << std::get<2>(E); continue; } // The vaiable has no definition - no need to add the entry. diff --git a/test/OpenMP/target_messages.cpp b/test/OpenMP/target_messages.cpp index 9bd8b3749e..5f22fbcaf6 100644 --- a/test/OpenMP/target_messages.cpp +++ b/test/OpenMP/target_messages.cpp @@ -13,14 +13,12 @@ // 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 incorrect: either the address or the ID is invalid. -// NO-REGION-NOT: Offloading entry for target region is incorrect: either the address or the ID is invalid. +// RUN: not %clang_cc1 -verify -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 #if defined(REGION_HOST) || defined(REGION_DEVICE) void foo() { #ifdef REGION_HOST -#pragma omp target +#pragma omp target // expected-error {{Offloading entry for target region in _Z3foov is incorrect: either the address or the ID is invalid.}} ; #endif #ifdef REGION_DEVICE