Skip to content

Commit 0dc410c

Browse files
format fix
1 parent d86188b commit 0dc410c

File tree

7 files changed

+43
-47
lines changed

7 files changed

+43
-47
lines changed

clang/lib/CodeGen/CGOpenMPRuntime.cpp

Lines changed: 34 additions & 39 deletions
Original file line numberDiff line numberDiff line change
@@ -1830,36 +1830,31 @@ void CGOpenMPRuntime::registerVTableOffloadEntry(llvm::GlobalVariable *VTable,
18301830
}
18311831

18321832
void CGOpenMPRuntime::emitAndRegisterVTable(CodeGenModule &CGM,
1833-
CXXRecordDecl *CXXRecord,
1834-
const VarDecl *VD) {
1835-
// Register C++ VTable to OpenMP Offload Entry if it's a new
1836-
// CXXRecordDecl.
1837-
if (CXXRecord && CXXRecord->isDynamicClass() &&
1838-
CGM.getOpenMPRuntime().VTableDeclMap.find(CXXRecord) ==
1839-
CGM.getOpenMPRuntime().VTableDeclMap.end()) {
1840-
CGM.getOpenMPRuntime().VTableDeclMap.try_emplace(CXXRecord, VD);
1841-
CGM.EmitVTable(CXXRecord);
1842-
CodeGenVTables VTables = CGM.getVTables();
1843-
llvm::GlobalVariable *VTablesAddr = VTables.GetAddrOfVTable(CXXRecord);
1844-
if (VTablesAddr)
1845-
CGM.getOpenMPRuntime().registerVTableOffloadEntry(VTablesAddr, VD);
1846-
// Emit VTable for all the fields containing dynamic CXXRecord
1847-
for (const FieldDecl *Field : CXXRecord->fields()) {
1848-
if (CXXRecordDecl *RecordDecl =
1849-
Field->getType()->getAsCXXRecordDecl())
1850-
emitAndRegisterVTable(CGM, RecordDecl, VD);
1851-
1852-
}
1853-
// Emit VTable for all dynamic parent class
1854-
for (CXXBaseSpecifier &Base : CXXRecord->bases()) {
1855-
if (CXXRecordDecl *BaseDecl =
1856-
Base.getType()->getAsCXXRecordDecl())
1857-
emitAndRegisterVTable(CGM, BaseDecl, VD);
1858-
1859-
}
1860-
}
1861-
};
1862-
1833+
CXXRecordDecl *CXXRecord,
1834+
const VarDecl *VD) {
1835+
// Register C++ VTable to OpenMP Offload Entry if it's a new
1836+
// CXXRecordDecl.
1837+
if (CXXRecord && CXXRecord->isDynamicClass() &&
1838+
CGM.getOpenMPRuntime().VTableDeclMap.find(CXXRecord) ==
1839+
CGM.getOpenMPRuntime().VTableDeclMap.end()) {
1840+
CGM.getOpenMPRuntime().VTableDeclMap.try_emplace(CXXRecord, VD);
1841+
CGM.EmitVTable(CXXRecord);
1842+
CodeGenVTables VTables = CGM.getVTables();
1843+
llvm::GlobalVariable *VTablesAddr = VTables.GetAddrOfVTable(CXXRecord);
1844+
if (VTablesAddr)
1845+
CGM.getOpenMPRuntime().registerVTableOffloadEntry(VTablesAddr, VD);
1846+
// Emit VTable for all the fields containing dynamic CXXRecord
1847+
for (const FieldDecl *Field : CXXRecord->fields()) {
1848+
if (CXXRecordDecl *RecordDecl = Field->getType()->getAsCXXRecordDecl())
1849+
emitAndRegisterVTable(CGM, RecordDecl, VD);
1850+
}
1851+
// Emit VTable for all dynamic parent class
1852+
for (CXXBaseSpecifier &Base : CXXRecord->bases()) {
1853+
if (CXXRecordDecl *BaseDecl = Base.getType()->getAsCXXRecordDecl())
1854+
emitAndRegisterVTable(CGM, BaseDecl, VD);
1855+
}
1856+
}
1857+
};
18631858

18641859
void CGOpenMPRuntime::registerVTable(const OMPExecutableDirective &D) {
18651860
// Register VTable by scanning through the map clause of OpenMP target region.
@@ -10070,15 +10065,15 @@ void CGOpenMPRuntime::scanForTargetRegionsFunctions(const Stmt *S,
1007010065
// Register vtable from device for target data and target directives.
1007110066
// Add this block here since scanForTargetRegionsFunctions ignores
1007210067
// target data by checking if S is a executable directive (target).
10073-
if (isa<OMPExecutableDirective>(S) &&
10074-
isOpenMPTargetDataManagementDirective(
10075-
dyn_cast<OMPExecutableDirective>(S)->getDirectiveKind())) {
10076-
auto &E = *dyn_cast<OMPExecutableDirective>(S);
10077-
// Don't need to check if it's device compile
10078-
// since scanForTargetRegionsFunctions currently only called
10079-
// in device compilation.
10080-
registerVTable(E);
10081-
}
10068+
if (isa<OMPExecutableDirective>(S) &&
10069+
isOpenMPTargetDataManagementDirective(
10070+
dyn_cast<OMPExecutableDirective>(S)->getDirectiveKind())) {
10071+
auto &E = *dyn_cast<OMPExecutableDirective>(S);
10072+
// Don't need to check if it's device compile
10073+
// since scanForTargetRegionsFunctions currently only called
10074+
// in device compilation.
10075+
registerVTable(E);
10076+
}
1008210077

1008310078
// Codegen OMP target directives that offload compute to the device.
1008410079
bool RequiresDeviceCodegen =

clang/lib/CodeGen/CodeGenModule.h

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -754,7 +754,7 @@ class CodeGenModule : public CodeGenTypeCache {
754754
// i32 @__isPlatformVersionAtLeast(i32, i32, i32, i32)
755755
llvm::FunctionCallee IsPlatformVersionAtLeastFn = nullptr;
756756

757-
// Store indirect CallExprs that are within an omp target region
757+
// Store indirect CallExprs that are within an omp target region
758758
llvm::SmallPtrSet<const CallExpr *, 16> OMPTargetCalls;
759759

760760
InstrProfStats &getPGOStats() { return PGOStats; }

clang/test/OpenMP/target_vtable_codegen_implicit_namespace.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -40,4 +40,4 @@ int main() {
4040
}
4141
}
4242
return 0;
43-
}
43+
}

clang/test/OpenMP/target_vtable_codegen_mult_inherritence.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -43,4 +43,4 @@ int main() {
4343
}
4444
}
4545
return 0;
46-
}
46+
}

llvm/lib/Frontend/OpenMP/OMPIRBuilder.cpp

Lines changed: 4 additions & 3 deletions
Original file line numberDiff line numberDiff line change
@@ -10249,8 +10249,8 @@ void OpenMPIRBuilder::createOffloadEntriesAndInfoMetadata(
1024910249
case OffloadEntriesInfoManager::OMPTargetGlobalVarEntryIndirect:
1025010250
case OffloadEntriesInfoManager::OMPTargetGlobalVarEntryIndirectVTable:
1025110251
if (!CE->getAddress()) {
10252-
ErrorFn(EMIT_MD_GLOBAL_VAR_INDIRECT_ERROR, E.second);
10253-
continue;
10252+
ErrorFn(EMIT_MD_GLOBAL_VAR_INDIRECT_ERROR, E.second);
10253+
continue;
1025410254
}
1025510255
break;
1025610256
default:
@@ -10702,7 +10702,8 @@ void OffloadEntriesInfoManager::registerDeviceGlobalVarEntryInfo(
1070210702
return;
1070310703
}
1070410704
if (Flags == OffloadEntriesInfoManager::OMPTargetGlobalVarEntryIndirect ||
10705-
Flags == OffloadEntriesInfoManager::OMPTargetGlobalVarEntryIndirectVTable)
10705+
Flags ==
10706+
OffloadEntriesInfoManager::OMPTargetGlobalVarEntryIndirectVTable)
1070610707
OffloadEntriesDeviceGlobalVar.try_emplace(VarName, OffloadingEntriesNum,
1070710708
Addr, VarSize, Flags, Linkage,
1070810709
VarName.str());

offload/include/omptarget.h

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -94,7 +94,7 @@ enum OpenMPOffloadingDeclareTargetFlags {
9494
OMP_DECLARE_TARGET_INDIRECT = 0x08,
9595
/// This is an entry corresponding to a requirement to be registered.
9696
OMP_REGISTER_REQUIRES = 0x10,
97-
/// Mark the entry global as being an indirect vtable.
97+
/// Mark the entry global as being an indirect vtable.
9898
OMP_DECLARE_TARGET_INDIRECT_VTABLE = 0x20,
9999
};
100100

offload/test/api/omp_indirect_call_table_manual.c

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -89,7 +89,7 @@ int main() {
8989
myClass obj_baz = {dispatchTable + 2};
9090
int aaa = 0;
9191

92-
#pragma omp target map(aaa) map (to: obj_foo, obj_bar, obj_baz)
92+
#pragma omp target map(aaa) map(to : obj_foo, obj_bar, obj_baz)
9393
{
9494
// Lookup
9595
fptr_t *foo_ptr = __llvm_omp_indirect_call_lookup(obj_foo.dispatchPtr);

0 commit comments

Comments
 (0)