[llvm] 2c31d5e - [OpenMP] Add IDs to OpenMP remarks
via llvm-commits
llvm-commits at lists.llvm.org
Fri Jul 16 11:07:23 PDT 2021
Author: Joseph Huber
Date: 2021-07-16T14:07:03-04:00
New Revision: 2c31d5ebfbbb36de37d41b6945d0c45efa512790
URL: https://github.com/llvm/llvm-project/commit/2c31d5ebfbbb36de37d41b6945d0c45efa512790
DIFF: https://github.com/llvm/llvm-project/commit/2c31d5ebfbbb36de37d41b6945d0c45efa512790.diff
LOG: [OpenMP] Add IDs to OpenMP remarks
This patch adds unique idenfitiers to the existing OpenMP remarks. This makes
it easier to identify the corresponding documentation for each remark that will
be hosted in the OpenMP webpage.
Depends on D105898
Reviewed By: jdoerfert
Differential Revision: https://reviews.llvm.org/D105939
Added:
Modified:
clang/test/OpenMP/remarks_parallel_in_multiple_target_state_machines.c
clang/test/OpenMP/remarks_parallel_in_target_state_machine.c
llvm/include/llvm/Transforms/IPO/Attributor.h
llvm/lib/Transforms/IPO/AttributorAttributes.cpp
llvm/lib/Transforms/IPO/OpenMPOpt.cpp
Removed:
################################################################################
diff --git a/clang/test/OpenMP/remarks_parallel_in_multiple_target_state_machines.c b/clang/test/OpenMP/remarks_parallel_in_multiple_target_state_machines.c
index 9a7693e744de0..3c74febc61874 100644
--- a/clang/test/OpenMP/remarks_parallel_in_multiple_target_state_machines.c
+++ b/clang/test/OpenMP/remarks_parallel_in_multiple_target_state_machines.c
@@ -8,23 +8,23 @@ void baz(void) __attribute__((assume("omp_no_openmp")));
void bar1(void) {
#pragma omp parallel // #0
- // safe-remark@#0 {{Parallel region is used in unknown ways. Will not attempt to rewrite the state machine.}}
+ // safe-remark@#0 {{Parallel region is used in unknown ways. Will not attempt to rewrite the state machine. [OMP101]}}
{
}
}
void bar2(void) {
#pragma omp parallel // #1
- // safe-remark@#1 {{Parallel region is used in unknown ways. Will not attempt to rewrite the state machine.}}
+ // safe-remark@#1 {{Parallel region is used in unknown ways. Will not attempt to rewrite the state machine. [OMP101]}}
{
}
}
void foo1(void) {
#pragma omp target teams // #2
- // all-remark@#2 {{Rewriting generic-mode kernel with a customized state machine.}}
+ // all-remark@#2 {{Rewriting generic-mode kernel with a customized state machine. [OMP131]}}
{
- baz(); // all-remark {{Value has potential side effects preventing SPMD-mode execution. Add `__attribute__((assume("ompx_spmd_amenable")))` to the called function to override.}}
+ baz(); // all-remark {{Value has potential side effects preventing SPMD-mode execution. Add `__attribute__((assume("ompx_spmd_amenable")))` to the called function to override. [OMP121]}}
#pragma omp parallel // #3
{
}
@@ -37,9 +37,9 @@ void foo1(void) {
void foo2(void) {
#pragma omp target teams // #5
- // all-remark@#5 {{Rewriting generic-mode kernel with a customized state machine.}}
+ // all-remark@#5 {{Rewriting generic-mode kernel with a customized state machine. [OMP131]}}
{
- baz(); // all-remark {{Value has potential side effects preventing SPMD-mode execution. Add `__attribute__((assume("ompx_spmd_amenable")))` to the called function to override.}}
+ baz(); // all-remark {{Value has potential side effects preventing SPMD-mode execution. Add `__attribute__((assume("ompx_spmd_amenable")))` to the called function to override. [OMP121]}}
#pragma omp parallel // #6
{
}
@@ -55,9 +55,9 @@ void foo2(void) {
void foo3(void) {
#pragma omp target teams // #8
- // all-remark@#8 {{Rewriting generic-mode kernel with a customized state machine.}}
+ // all-remark@#8 {{Rewriting generic-mode kernel with a customized state machine. [OMP131]}}
{
- baz(); // all-remark {{Value has potential side effects preventing SPMD-mode execution. Add `__attribute__((assume("ompx_spmd_amenable")))` to the called function to override.}}
+ baz(); // all-remark {{Value has potential side effects preventing SPMD-mode execution. Add `__attribute__((assume("ompx_spmd_amenable")))` to the called function to override. [OMP121]}}
#pragma omp parallel // #9
{
}
@@ -83,4 +83,4 @@ void spmd(void) {
}
}
-// all-remark@* 9 {{OpenMP runtime call __kmpc_global_thread_num deduplicated}}
+// all-remark@* 9 {{OpenMP runtime call __kmpc_global_thread_num deduplicated. [OMP170]}}
diff --git a/clang/test/OpenMP/remarks_parallel_in_target_state_machine.c b/clang/test/OpenMP/remarks_parallel_in_target_state_machine.c
index 69891d1a36384..38c8a80502599 100644
--- a/clang/test/OpenMP/remarks_parallel_in_target_state_machine.c
+++ b/clang/test/OpenMP/remarks_parallel_in_target_state_machine.c
@@ -8,16 +8,16 @@ void baz(void) __attribute__((assume("omp_no_openmp")));
void bar(void) {
#pragma omp parallel // #1 \
- // expected-remark@#1 {{Parallel region is used in unknown ways. Will not attempt to rewrite the state machine.}}
+ // expected-remark@#1 {{Parallel region is used in unknown ways. Will not attempt to rewrite the state machine. [OMP101]}}
{
}
}
void foo(void) {
#pragma omp target teams // #2
- // expected-remark@#2 {{Rewriting generic-mode kernel with a customized state machine.}}
+ // expected-remark@#2 {{Rewriting generic-mode kernel with a customized state machine. [OMP131]}}
{
- baz(); // expected-remark {{Value has potential side effects preventing SPMD-mode execution. Add `__attribute__((assume("ompx_spmd_amenable")))` to the called function to override.}}
+ baz(); // expected-remark {{Value has potential side effects preventing SPMD-mode execution. Add `__attribute__((assume("ompx_spmd_amenable")))` to the called function to override. [OMP121]}}
#pragma omp parallel
{
}
@@ -40,4 +40,4 @@ void spmd(void) {
}
}
-// expected-remark@* {{OpenMP runtime call __kmpc_global_thread_num deduplicated}}
+// expected-remark@* {{OpenMP runtime call __kmpc_global_thread_num deduplicated. [OMP170]}}
diff --git a/llvm/include/llvm/Transforms/IPO/Attributor.h b/llvm/include/llvm/Transforms/IPO/Attributor.h
index bf117791b499d..6829f6c231744 100644
--- a/llvm/include/llvm/Transforms/IPO/Attributor.h
+++ b/llvm/include/llvm/Transforms/IPO/Attributor.h
@@ -1587,7 +1587,13 @@ struct Attributor {
Function *F = I->getFunction();
auto &ORE = OREGetter.getValue()(F);
- ORE.emit([&]() { return RemarkCB(RemarkKind(PassName, RemarkName, I)); });
+ if (RemarkName.startswith("OMP"))
+ ORE.emit([&]() {
+ return RemarkCB(RemarkKind(PassName, RemarkName, I))
+ << " [" << RemarkName << "]";
+ });
+ else
+ ORE.emit([&]() { return RemarkCB(RemarkKind(PassName, RemarkName, I)); });
}
/// Emit a remark on a function.
@@ -1599,7 +1605,13 @@ struct Attributor {
auto &ORE = OREGetter.getValue()(F);
- ORE.emit([&]() { return RemarkCB(RemarkKind(PassName, RemarkName, F)); });
+ if (RemarkName.startswith("OMP"))
+ ORE.emit([&]() {
+ return RemarkCB(RemarkKind(PassName, RemarkName, F))
+ << " [" << RemarkName << "]";
+ });
+ else
+ ORE.emit([&]() { return RemarkCB(RemarkKind(PassName, RemarkName, F)); });
}
/// Helper struct used in the communication between an abstract attribute (AA)
diff --git a/llvm/lib/Transforms/IPO/AttributorAttributes.cpp b/llvm/lib/Transforms/IPO/AttributorAttributes.cpp
index ca629a4df4b82..92edc684d3a35 100644
--- a/llvm/lib/Transforms/IPO/AttributorAttributes.cpp
+++ b/llvm/lib/Transforms/IPO/AttributorAttributes.cpp
@@ -5039,7 +5039,10 @@ struct AAHeapToStackFunction final : public AAHeapToStack {
return OR << "Moving globalized variable to the stack.";
return OR << "Moving memory allocation from the heap to the stack.";
};
- A.emitRemark<OptimizationRemark>(AI.CB, "HeapToStack", Remark);
+ if (AI.LibraryFunctionId == LibFunc___kmpc_alloc_shared)
+ A.emitRemark<OptimizationRemark>(AI.CB, "OMP110", Remark);
+ else
+ A.emitRemark<OptimizationRemark>(AI.CB, "HeapToStack", Remark);
Value *Size;
Optional<APInt> SizeAPI = getSize(A, *this, AI);
@@ -5335,8 +5338,7 @@ ChangeStatus AAHeapToStackFunction::updateImpl(Attributor &A) {
};
if (AI.LibraryFunctionId == LibFunc___kmpc_alloc_shared)
- A.emitRemark<OptimizationRemarkMissed>(AI.CB, "HeapToStackFailed",
- Remark);
+ A.emitRemark<OptimizationRemarkMissed>(AI.CB, "OMP113", Remark);
LLVM_DEBUG(dbgs() << "[H2S] Bad user: " << *UserI << "\n");
ValidUsesOnly = false;
diff --git a/llvm/lib/Transforms/IPO/OpenMPOpt.cpp b/llvm/lib/Transforms/IPO/OpenMPOpt.cpp
index 3cb7c17534bf5..eab09dd499ba1 100644
--- a/llvm/lib/Transforms/IPO/OpenMPOpt.cpp
+++ b/llvm/lib/Transforms/IPO/OpenMPOpt.cpp
@@ -949,8 +949,7 @@ struct OpenMPOpt {
return OR << ".";
};
- emitRemark<OptimizationRemark>(MergableCIs.front(),
- "OpenMPParallelRegionMerging", Remark);
+ emitRemark<OptimizationRemark>(MergableCIs.front(), "OMP150", Remark);
Function *OriginalFn = BB->getParent();
LLVM_DEBUG(dbgs() << TAG << "Merge " << MergableCIs.size()
@@ -1204,8 +1203,7 @@ struct OpenMPOpt {
auto Remark = [&](OptimizationRemark OR) {
return OR << "Removing parallel region with no side-effects.";
};
- emitRemark<OptimizationRemark>(CI, "OpenMPParallelRegionDeletion",
- Remark);
+ emitRemark<OptimizationRemark>(CI, "OMP160", Remark);
CGUpdater.removeCallSite(*CI);
CI->eraseFromParent();
@@ -1311,7 +1309,7 @@ struct OpenMPOpt {
<< "Found thread data sharing on the GPU. "
<< "Expect degraded performance due to data globalization.";
};
- emitRemark<OptimizationRemarkMissed>(CI, "OpenMPGlobalization", Remark);
+ emitRemark<OptimizationRemarkMissed>(CI, "OMP112", Remark);
}
return false;
@@ -1593,9 +1591,9 @@ struct OpenMPOpt {
<< ore::NV("OpenMPOptRuntime", RFI.Name) << " deduplicated.";
};
if (CI->getDebugLoc())
- emitRemark<OptimizationRemark>(CI, "OpenMPRuntimeDeduplicated", Remark);
+ emitRemark<OptimizationRemark>(CI, "OMP170", Remark);
else
- emitRemark<OptimizationRemark>(&F, "OpenMPRuntimeDeduplicated", Remark);
+ emitRemark<OptimizationRemark>(&F, "OMP170", Remark);
CGUpdater.removeCallSite(*CI);
CI->replaceAllUsesWith(ReplVal);
@@ -1702,7 +1700,14 @@ struct OpenMPOpt {
Function *F = I->getParent()->getParent();
auto &ORE = OREGetter(F);
- ORE.emit([&]() { return RemarkCB(RemarkKind(DEBUG_TYPE, RemarkName, I)); });
+ if (RemarkName.startswith("OMP"))
+ ORE.emit([&]() {
+ return RemarkCB(RemarkKind(DEBUG_TYPE, RemarkName, I))
+ << " [" << RemarkName << "]";
+ });
+ else
+ ORE.emit(
+ [&]() { return RemarkCB(RemarkKind(DEBUG_TYPE, RemarkName, I)); });
}
/// Emit a remark on a function.
@@ -1711,7 +1716,14 @@ struct OpenMPOpt {
RemarkCallBack &&RemarkCB) const {
auto &ORE = OREGetter(F);
- ORE.emit([&]() { return RemarkCB(RemarkKind(DEBUG_TYPE, RemarkName, F)); });
+ if (RemarkName.startswith("OMP"))
+ ORE.emit([&]() {
+ return RemarkCB(RemarkKind(DEBUG_TYPE, RemarkName, F))
+ << " [" << RemarkName << "]";
+ });
+ else
+ ORE.emit(
+ [&]() { return RemarkCB(RemarkKind(DEBUG_TYPE, RemarkName, F)); });
}
/// The underlying module.
@@ -1880,8 +1892,7 @@ bool OpenMPOpt::rewriteDeviceCodeStateMachine() {
<< (UnknownUse ? "unknown" : "unexpected")
<< " ways. Will not attempt to rewrite the state machine.";
};
- emitRemark<OptimizationRemarkAnalysis>(F, "OpenMPParallelRegionInNonSPMD",
- Remark);
+ emitRemark<OptimizationRemarkAnalysis>(F, "OMP101", Remark);
continue;
}
@@ -1893,8 +1904,7 @@ bool OpenMPOpt::rewriteDeviceCodeStateMachine() {
return ORA << "Parallel region is not called from a unique kernel. "
"Will not attempt to rewrite the state machine.";
};
- emitRemark<OptimizationRemarkAnalysis>(
- F, "OpenMPParallelRegionInMultipleKernesl", Remark);
+ emitRemark<OptimizationRemarkAnalysis>(F, "OMP102", Remark);
continue;
}
@@ -2581,8 +2591,7 @@ struct AAHeapToSharedFunction : public AAHeapToShared {
<< ((AllocSize->getZExtValue() != 1) ? " bytes " : " byte ")
<< "of shared memory.";
};
- A.emitRemark<OptimizationRemark>(CB, "OpenMPReplaceGlobalization",
- Remark);
+ A.emitRemark<OptimizationRemark>(CB, "OMP111", Remark);
SharedMem->setAlignment(MaybeAlign(32));
@@ -2823,8 +2832,8 @@ struct AAKernelInfoFunction : AAKernelInfo {
}
return ORA << ".";
};
- A.emitRemark<OptimizationRemarkAnalysis>(
- NonCompatibleI, "OpenMPKernelNonSPMDMode", Remark);
+ A.emitRemark<OptimizationRemarkAnalysis>(NonCompatibleI, "OMP121",
+ Remark);
LLVM_DEBUG(dbgs() << TAG << "SPMD-incompatible side-effect: "
<< *NonCompatibleI << "\n");
@@ -2864,8 +2873,7 @@ struct AAKernelInfoFunction : AAKernelInfo {
auto Remark = [&](OptimizationRemark OR) {
return OR << "Transformed generic-mode kernel to SPMD-mode.";
};
- A.emitRemark<OptimizationRemark>(KernelInitCB, "OpenMPKernelSPMDMode",
- Remark);
+ A.emitRemark<OptimizationRemark>(KernelInitCB, "OMP120", Remark);
return true;
};
@@ -2909,8 +2917,7 @@ struct AAKernelInfoFunction : AAKernelInfo {
auto Remark = [&](OptimizationRemark OR) {
return OR << "Removing unused state machine from generic-mode kernel.";
};
- A.emitRemark<OptimizationRemark>(
- KernelInitCB, "OpenMPKernelWithoutStateMachine", Remark);
+ A.emitRemark<OptimizationRemark>(KernelInitCB, "OMP130", Remark);
return ChangeStatus::CHANGED;
}
@@ -2923,8 +2930,7 @@ struct AAKernelInfoFunction : AAKernelInfo {
return OR << "Rewriting generic-mode kernel with a customized state "
"machine.";
};
- A.emitRemark<OptimizationRemark>(
- KernelInitCB, "OpenMPKernelWithCustomizedStateMachine", Remark);
+ A.emitRemark<OptimizationRemark>(KernelInitCB, "OMP131", Remark);
} else {
++NumOpenMPTargetRegionKernelsCustomStateMachineWithFallback;
@@ -2932,9 +2938,7 @@ struct AAKernelInfoFunction : AAKernelInfo {
return OR << "Generic-mode kernel is executed with a customized state "
"machine that requires a fallback.";
};
- A.emitRemark<OptimizationRemarkAnalysis>(
- KernelInitCB, "OpenMPKernelWithCustomizedStateMachineAndFallback",
- Remark);
+ A.emitRemark<OptimizationRemarkAnalysis>(KernelInitCB, "OMP132", Remark);
// Tell the user why we ended up with a fallback.
for (CallBase *UnknownParallelRegionCB : ReachedUnknownParallelRegions) {
@@ -2945,9 +2949,8 @@ struct AAKernelInfoFunction : AAKernelInfo {
<< "`__attribute__((assume(\"omp_no_parallelism\")))` to "
"override.";
};
- A.emitRemark<OptimizationRemarkAnalysis>(
- UnknownParallelRegionCB,
- "OpenMPKernelWithCustomizedStateMachineAndFallback", Remark);
+ A.emitRemark<OptimizationRemarkAnalysis>(UnknownParallelRegionCB,
+ "OMP133", Remark);
}
}
@@ -3741,7 +3744,7 @@ PreservedAnalyses OpenMPOptPass::run(Module &M, ModuleAnalysisManager &AM) {
auto EmitRemark = [&](Function &F) {
auto &ORE = FAM.getResult<OptimizationRemarkEmitterAnalysis>(F);
ORE.emit([&]() {
- OptimizationRemarkAnalysis ORA(DEBUG_TYPE, "InternalizationFailure", &F);
+ OptimizationRemarkAnalysis ORA(DEBUG_TYPE, "OMP140", &F);
return ORA << "Could not internalize function. "
<< "Some optimizations may not be possible.";
});
More information about the llvm-commits
mailing list