jhuber6 updated this revision to Diff 358461.
jhuber6 added a comment.
Herald added a project: clang.
Herald added a subscriber: cfe-commits.
Adding test case for IDs.
Repository:
rG LLVM Github Monorepo
CHANGES SINCE LAST ACTION
https://reviews.llvm.org/D105939/new/
https://reviews.llvm.org/D105939
Files:
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
Index: llvm/lib/Transforms/IPO/OpenMPOpt.cpp
===================================================================
--- llvm/lib/Transforms/IPO/OpenMPOpt.cpp
+++ llvm/lib/Transforms/IPO/OpenMPOpt.cpp
@@ -943,8 +943,7 @@
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()
@@ -1040,8 +1039,7 @@
<< ".";
};
if (CI != MergableCIs.front())
- emitRemark<OptimizationRemark>(CI, "OpenMPParallelRegionMerging",
- Remark);
+ emitRemark<OptimizationRemark>(CI, "OMP150", Remark);
CI->eraseFromParent();
}
@@ -1210,8 +1208,7 @@
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();
@@ -1317,7 +1314,7 @@
<< "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;
@@ -1598,7 +1595,7 @@
return OR << "OpenMP runtime call "
<< ore::NV("OpenMPOptRuntime", RFI.Name) << " deduplicated.";
};
- emitRemark<OptimizationRemark>(&F, "OpenMPRuntimeDeduplicated", Remark);
+ emitRemark<OptimizationRemark>(&F, "OMP170", Remark);
CGUpdater.removeCallSite(*CI);
CI->replaceAllUsesWith(ReplVal);
@@ -1705,7 +1702,14 @@
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.
@@ -1714,7 +1718,14 @@
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.
@@ -1883,8 +1894,7 @@
<< (UnknownUse ? "unknown" : "unexpected")
<< " ways. Will not attempt to rewrite the state machine.";
};
- emitRemark<OptimizationRemarkAnalysis>(F, "OpenMPParallelRegionInNonSPMD",
- Remark);
+ emitRemark<OptimizationRemarkAnalysis>(F, "OMP101", Remark);
continue;
}
@@ -1896,8 +1906,7 @@
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;
}
@@ -2584,8 +2593,7 @@
<< ((AllocSize->getZExtValue() != 1) ? " bytes " : " byte ")
<< "of shared memory";
};
- A.emitRemark<OptimizationRemark>(CB, "OpenMPReplaceGlobalization",
- Remark);
+ A.emitRemark<OptimizationRemark>(CB, "OMP111", Remark);
SharedMem->setAlignment(MaybeAlign(32));
@@ -2816,14 +2824,14 @@
auto Remark = [&](OptimizationRemarkAnalysis ORA) {
ORA << "Value has potential side effects preventing SPMD-mode "
"execution";
- if (auto *CI = dyn_cast<CallBase>(NonCompatibleI)) {
+ if (isa<CallBase>(NonCompatibleI)) {
ORA << ". Add `__attribute__((assume(\"ompx_spmd_amenable\")))` to "
"the called function to override";
}
return ORA << ".";
};
- A.emitRemark<OptimizationRemarkAnalysis>(
- NonCompatibleI, "OpenMPKernelNonSPMDMode", Remark);
+ A.emitRemark<OptimizationRemarkAnalysis>(NonCompatibleI, "OMP121",
+ Remark);
LLVM_DEBUG(dbgs() << TAG << "SPMD-incompatible side-effect: "
<< *NonCompatibleI << "\n");
@@ -2863,8 +2871,7 @@
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;
};
@@ -2908,8 +2915,7 @@
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;
}
@@ -2922,8 +2928,7 @@
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;
@@ -2937,9 +2942,7 @@
ReachedUnknownParallelRegions.size())
<< " unkown parallel regions].";
};
- 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) {
@@ -2950,9 +2953,8 @@
<< "`__attribute__((assume(\"omp_no_parallelism\")))` to "
"override.";
};
- A.emitRemark<OptimizationRemarkAnalysis>(
- UnknownParallelRegionCB,
- "OpenMPKernelWithCustomizedStateMachineAndFallback", Remark);
+ A.emitRemark<OptimizationRemarkAnalysis>(UnknownParallelRegionCB,
+ "OMP133", Remark);
}
}
@@ -3501,7 +3503,7 @@
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.";
});
Index: llvm/lib/Transforms/IPO/AttributorAttributes.cpp
===================================================================
--- llvm/lib/Transforms/IPO/AttributorAttributes.cpp
+++ llvm/lib/Transforms/IPO/AttributorAttributes.cpp
@@ -5039,7 +5039,10 @@
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 @@
};
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;
Index: llvm/include/llvm/Transforms/IPO/Attributor.h
===================================================================
--- llvm/include/llvm/Transforms/IPO/Attributor.h
+++ llvm/include/llvm/Transforms/IPO/Attributor.h
@@ -1585,7 +1585,13 @@
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.
@@ -1597,7 +1603,13 @@
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)
Index: clang/test/OpenMP/remarks_parallel_in_target_state_machine.c
===================================================================
--- clang/test/OpenMP/remarks_parallel_in_target_state_machine.c
+++ clang/test/OpenMP/remarks_parallel_in_target_state_machine.c
@@ -8,16 +8,16 @@
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 @@
}
}
-// expected-remark@* {{OpenMP runtime call __kmpc_global_thread_num deduplicated}}
+// expected-remark@* {{OpenMP runtime call __kmpc_global_thread_num deduplicated. [OMP170]}}
Index: clang/test/OpenMP/remarks_parallel_in_multiple_target_state_machines.c
===================================================================
--- clang/test/OpenMP/remarks_parallel_in_multiple_target_state_machines.c
+++ clang/test/OpenMP/remarks_parallel_in_multiple_target_state_machines.c
@@ -8,23 +8,23 @@
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 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 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 @@
}
}
-// all-remark@* 9 {{OpenMP runtime call __kmpc_global_thread_num deduplicated}}
+// all-remark@* 9 {{OpenMP runtime call __kmpc_global_thread_num deduplicated. [OMP170]}}
_______________________________________________
cfe-commits mailing list
cfe-commits@lists.llvm.org
https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
