maxjhandsome
/
llvm-project
forked from OSchip/llvm-project
1
0
Fork 0
Code Issues Pull Requests Packages Releases Wiki Activity

[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
This commit is contained in:
Joseph Huber 2021-07-13 15:31:44 -04:00 committed by Huber, Joseph
parent eef6601b0f
commit 2c31d5ebfb
5 changed files with 65 additions and 48 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

18
clang/test/OpenMP/remarks_parallel_in_multiple_target_state_machines.c
View File

@ -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]}}

8
clang/test/OpenMP/remarks_parallel_in_target_state_machine.c
View File

@ -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]}}

16
llvm/include/llvm/Transforms/IPO/Attributor.h
View File

@ -1587,7 +1587,13 @@ public:
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 @@ public:
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)

8
llvm/lib/Transforms/IPO/AttributorAttributes.cpp
View File

@ -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;

63
llvm/lib/Transforms/IPO/OpenMPOpt.cpp
View File

@ -949,8 +949,7 @@ private:
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 @@ private:
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 @@ private:
<< "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 @@ private:
<< 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 @@ private:
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 @@ private:
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.";
});