This is an archive of the discontinued LLVM Phabricator instance.

Differential D91552

[OPENMP]Fix PR48076: mapping of data member pointer.
ClosedPublic

Authored by ABataev on Nov 16 2020, 10:22 AM.

Details

Reviewers
jdoerfert
jhuber6
Commits
rG5292187a2d62: [OPENMP]Fix PR48076: mapping of data member pointer.
Summary

If the data member pointer is mapped, the compiler tries to optimize the
mapping of such data by discarding explicit mapping flags and trying to
emit combined data instead. In some cases, this optimization is not
quite correctly implemented and it leads to a program crash at the
runtime. Instead, if the data member is mapped, just emit it as is and
do not emit combined mapping flags for it.

Diff Detail

Event Timeline

ABataev created this revision.Nov 16 2020, 10:22 AM
Herald added a project: Restricted Project. · View Herald TranscriptNov 16 2020, 10:22 AM
Herald added subscribers: guansong, yaxunl. · View Herald Transcript
ABataev requested review of this revision.Nov 16 2020, 10:22 AM
ABataev updated this revision to Diff 305553.

Rename

ABataev retitled this revision from [OPENMP]Fix mapping of data member pointer. to [OPENMP]Fix PR48076: mapping of data member pointer..Nov 16 2020, 10:24 AM
Harbormaster completed remote builds in B78982: Diff 305553.Nov 16 2020, 10:54 AM
Harbormaster completed remote builds in B78981: Diff 305552.Nov 16 2020, 10:58 AM
jhuber6 accepted this revision.Nov 16 2020, 11:17 AM

LGTM, fixed the problem on my machine and passes the unit tests.

This revision is now accepted and ready to land.Nov 16 2020, 11:17 AM
This revision was landed with ongoing or failed builds.Nov 17 2020, 7:27 AM
Closed by commit rG5292187a2d62: [OPENMP]Fix PR48076: mapping of data member pointer. (authored by ABataev). · Explain Why
This revision was automatically updated to reflect the committed changes.
ABataev added a commit: rG5292187a2d62: [OPENMP]Fix PR48076: mapping of data member pointer..

Revision Contents

Diff 305790

clang/lib/CodeGen/CGOpenMPRuntime.cpp

  • This file is larger than 256 KB, so syntax highlighting is disabled by default.
Show First 20 Lines • Show All 7,120 Lines • ▼ Show 20 Linespublic:
/// [ValueDecl *] --> {LE(FieldIndex, Pointer), /// [ValueDecl *] --> {LE(FieldIndex, Pointer),
/// HE(FieldIndex, Pointer)} /// HE(FieldIndex, Pointer)}
struct StructRangeInfoTy { struct StructRangeInfoTy {
std::pair<unsigned /*FieldIndex*/, Address /*Pointer*/> LowestElem = { std::pair<unsigned /*FieldIndex*/, Address /*Pointer*/> LowestElem = {
0, Address::invalid()}; 0, Address::invalid()};
std::pair<unsigned /*FieldIndex*/, Address /*Pointer*/> HighestElem = { std::pair<unsigned /*FieldIndex*/, Address /*Pointer*/> HighestElem = {
0, Address::invalid()}; 0, Address::invalid()};
Address Base = Address::invalid(); Address Base = Address::invalid();
bool IsArraySection = false;
}; };
private: private:
/// Kind that defines how a device pointer has to be returned. /// Kind that defines how a device pointer has to be returned.
struct MapInfo { struct MapInfo {
OMPClauseMappableExprCommon::MappableExprComponentListRef Components; OMPClauseMappableExprCommon::MappableExprComponentListRef Components;
OpenMPMapClauseKind MapType = OMPC_MAP_unknown; OpenMPMapClauseKind MapType = OMPC_MAP_unknown;
ArrayRef<OpenMPMapModifierKind> MapModifiers; ArrayRef<OpenMPMapModifierKind> MapModifiers;
▲ Show 20 Lines • Show All 643 Lines • ▼ Show 20 Lines for (; I != CE; ++I) {
CGF.Builder.CreateIntCast(Size, CGF.Int64Ty, /*isSigned=*/true)); CGF.Builder.CreateIntCast(Size, CGF.Int64Ty, /*isSigned=*/true));
CombinedInfo.Types.push_back(Flags); CombinedInfo.Types.push_back(Flags);
CombinedInfo.Mappers.push_back(nullptr); CombinedInfo.Mappers.push_back(nullptr);
CombinedInfo.NonContigInfo.Dims.push_back(IsNonContiguous ? DimSize CombinedInfo.NonContigInfo.Dims.push_back(IsNonContiguous ? DimSize
: 1); : 1);
break; break;
} }
llvm::Value *Size = getExprTypeSize(I->getAssociatedExpression()); llvm::Value *Size = getExprTypeSize(I->getAssociatedExpression());
if (!IsMemberPointerOrAddr) { if (!IsMemberPointerOrAddr ||
(Next == CE && MapType != OMPC_MAP_unknown)) {
CombinedInfo.BasePointers.push_back(BP.getPointer()); CombinedInfo.BasePointers.push_back(BP.getPointer());
CombinedInfo.Pointers.push_back(LB.getPointer()); CombinedInfo.Pointers.push_back(LB.getPointer());
CombinedInfo.Sizes.push_back( CombinedInfo.Sizes.push_back(
CGF.Builder.CreateIntCast(Size, CGF.Int64Ty, /*isSigned=*/true)); CGF.Builder.CreateIntCast(Size, CGF.Int64Ty, /*isSigned=*/true));
CombinedInfo.NonContigInfo.Dims.push_back(IsNonContiguous ? DimSize CombinedInfo.NonContigInfo.Dims.push_back(IsNonContiguous ? DimSize
: 1); : 1);
// If Mapper is valid, the last component inherits the mapper. // If Mapper is valid, the last component inherits the mapper.
▲ Show 20 Lines • Show All 51 Lines • ▼ Show 20 Lines for (; I != CE; ++I) {
PartialStruct.Base = BP; PartialStruct.Base = BP;
} else if (FieldIndex < PartialStruct.LowestElem.first) { } else if (FieldIndex < PartialStruct.LowestElem.first) {
PartialStruct.LowestElem = {FieldIndex, LB}; PartialStruct.LowestElem = {FieldIndex, LB};
} else if (FieldIndex > PartialStruct.HighestElem.first) { } else if (FieldIndex > PartialStruct.HighestElem.first) {
PartialStruct.HighestElem = {FieldIndex, LB}; PartialStruct.HighestElem = {FieldIndex, LB};
} }
} }
// Need to emit combined struct for array sections.
if (IsFinalArraySection || IsNonContiguous)
PartialStruct.IsArraySection = true;
// If we have a final array section, we are done with this expression. // If we have a final array section, we are done with this expression.
if (IsFinalArraySection) if (IsFinalArraySection)
break; break;
// The pointer becomes the base for the next element. // The pointer becomes the base for the next element.
if (Next != CE) if (Next != CE)
BP = LB; BP = LB;
▲ Show 20 Lines • Show All 319 Lines • ▼ Show 20 Linespublic:
/// Generate code for the combined entry if we have a partially mapped struct /// Generate code for the combined entry if we have a partially mapped struct
/// and take care of the mapping flags of the arguments corresponding to /// and take care of the mapping flags of the arguments corresponding to
/// individual struct members. /// individual struct members.
void emitCombinedEntry(MapCombinedInfoTy &CombinedInfo, void emitCombinedEntry(MapCombinedInfoTy &CombinedInfo,
MapFlagsArrayTy &CurTypes, MapFlagsArrayTy &CurTypes,
const StructRangeInfoTy &PartialStruct, const StructRangeInfoTy &PartialStruct,
bool NotTargetParams = false) const { bool NotTargetParams = false) const {
if (CurTypes.size() == 1 &&
((CurTypes.back() & OMP_MAP_MEMBER_OF) != OMP_MAP_MEMBER_OF) &&
!PartialStruct.IsArraySection)
return;
// Base is the base of the struct // Base is the base of the struct
CombinedInfo.BasePointers.push_back(PartialStruct.Base.getPointer()); CombinedInfo.BasePointers.push_back(PartialStruct.Base.getPointer());
// Pointer is the address of the lowest element // Pointer is the address of the lowest element
llvm::Value *LB = PartialStruct.LowestElem.second.getPointer(); llvm::Value *LB = PartialStruct.LowestElem.second.getPointer();
CombinedInfo.Pointers.push_back(LB); CombinedInfo.Pointers.push_back(LB);
// There should not be a mapper for a combined entry. // There should not be a mapper for a combined entry.
CombinedInfo.Mappers.push_back(nullptr); CombinedInfo.Mappers.push_back(nullptr);
// Size is (addr of {highest+1} element) - (addr of lowest element) // Size is (addr of {highest+1} element) - (addr of lowest element)
▲ Show 20 Lines • Show All 1,734 Lines • ▼ Show 20 Lines for (CapturedStmt::const_capture_iterator CI = CS.capture_begin(),
} else { } else {
// If we have any information in the map clause, we use it, otherwise we // If we have any information in the map clause, we use it, otherwise we
// just do a default mapping. // just do a default mapping.
MEHandler.generateInfoForCapture(CI, *CV, CurInfo, PartialStruct); MEHandler.generateInfoForCapture(CI, *CV, CurInfo, PartialStruct);
if (!CI->capturesThis()) if (!CI->capturesThis())
MappedVarSet.insert(CI->getCapturedVar()); MappedVarSet.insert(CI->getCapturedVar());
else else
MappedVarSet.insert(nullptr); MappedVarSet.insert(nullptr);
if (CurInfo.BasePointers.empty()) if (CurInfo.BasePointers.empty() && !PartialStruct.Base.isValid())
MEHandler.generateDefaultMapInfo(*CI, **RI, *CV, CurInfo); MEHandler.generateDefaultMapInfo(*CI, **RI, *CV, CurInfo);
// Generate correct mapping for variables captured by reference in // Generate correct mapping for variables captured by reference in
// lambdas. // lambdas.
if (CI->capturesVariable()) if (CI->capturesVariable())
MEHandler.generateInfoForLambdaCaptures(CI->getCapturedVar(), *CV, MEHandler.generateInfoForLambdaCaptures(CI->getCapturedVar(), *CV,
CurInfo, LambdaPointers); CurInfo, LambdaPointers);
} }
// We expect to have at least an element of information for this capture. // We expect to have at least an element of information for this capture.
assert(!CurInfo.BasePointers.empty() && assert((!CurInfo.BasePointers.empty() || PartialStruct.Base.isValid()) &&
"Non-existing map pointer for capture!"); "Non-existing map pointer for capture!");
assert(CurInfo.BasePointers.size() == CurInfo.Pointers.size() && assert(CurInfo.BasePointers.size() == CurInfo.Pointers.size() &&
CurInfo.BasePointers.size() == CurInfo.Sizes.size() && CurInfo.BasePointers.size() == CurInfo.Sizes.size() &&
CurInfo.BasePointers.size() == CurInfo.Types.size() && CurInfo.BasePointers.size() == CurInfo.Types.size() &&
CurInfo.BasePointers.size() == CurInfo.Mappers.size() && CurInfo.BasePointers.size() == CurInfo.Mappers.size() &&
"Inconsistent map information sizes!"); "Inconsistent map information sizes!");
// If there is an entry in PartialStruct it means we have a struct with // If there is an entry in PartialStruct it means we have a struct with
▲ Show 20 Lines • Show All 2,548 Lines • Show Last 20 Lines

clang/test/OpenMP/declare_mapper_codegen.cpp

Show First 20 Lines • Show All 680 Lines • ▼ Show 20 Lines
// CK1: [[LHEAD]] // CK1: [[LHEAD]]
// CK1: [[ISEMPTY:%.+]] = icmp eq %class.C** [[PTRBEGIN]], [[PTREND]] // CK1: [[ISEMPTY:%.+]] = icmp eq %class.C** [[PTRBEGIN]], [[PTREND]]
// CK1: br i1 [[ISEMPTY]], label %[[DONE:[^,]+]], label %[[LBODY:[^,]+]] // CK1: br i1 [[ISEMPTY]], label %[[DONE:[^,]+]], label %[[LBODY:[^,]+]]
// CK1: [[LBODY]] // CK1: [[LBODY]]
// CK1: [[PTR:%.+]] = phi %class.C** [ [[PTRBEGIN]], %[[LHEAD]] ], [ [[PTRNEXT:%.+]], %[[LCORRECT:[^,]+]] ] // CK1: [[PTR:%.+]] = phi %class.C** [ [[PTRBEGIN]], %[[LHEAD]] ], [ [[PTRNEXT:%.+]], %[[LCORRECT:[^,]+]] ]
// CK1: [[OBJ:%.+]] = load %class.C*, %class.C** [[PTR]] // CK1: [[OBJ:%.+]] = load %class.C*, %class.C** [[PTR]]
// CK1-DAG: [[ABEGIN:%.+]] = getelementptr inbounds %class.C, %class.C* [[OBJ]], i32 0, i32 0 // CK1-DAG: [[ABEGIN:%.+]] = getelementptr inbounds %class.C, %class.C* [[OBJ]], i32 0, i32 0
// CK1-DAG: [[AEND:%.+]] = getelementptr i32, i32* [[ABEGIN]], i32 1
// CK1-DAG: [[ABEGINV:%.+]] = bitcast i32* [[ABEGIN]] to i8*
// CK1-DAG: [[AENDV:%.+]] = bitcast i32* [[AEND]] to i8*
// CK1-DAG: [[ABEGINI:%.+]] = ptrtoint i8* [[ABEGINV]] to i64
// CK1-DAG: [[AENDI:%.+]] = ptrtoint i8* [[AENDV]] to i64
// CK1-DAG: [[CSIZE:%.+]] = sub i64 [[AENDI]], [[ABEGINI]]
// CK1-DAG: [[CUSIZE:%.+]] = sdiv exact i64 [[CSIZE]], ptrtoint (i8* getelementptr (i8, i8* null, i32 1) to i64)
// CK1-DAG: [[BPTRADDR0BC:%.+]] = bitcast %class.C* [[OBJ]] to i8*
// CK1-DAG: [[PTRADDR0BC:%.+]] = bitcast i32* [[ABEGIN]] to i8*
// CK1-DAG: [[PRESIZE:%.+]] = call i64 @__tgt_mapper_num_components(i8* [[HANDLE]]) // CK1-DAG: [[PRESIZE:%.+]] = call i64 @__tgt_mapper_num_components(i8* [[HANDLE]])
// CK1-DAG: [[SHIPRESIZE:%.+]] = shl i64 [[PRESIZE]], 48 // CK1-DAG: [[SHIPRESIZE:%.+]] = shl i64 [[PRESIZE]], 48
// CK1-DAG: br label %[[MEMBER:[^,]+]]
// CK1-DAG: [[MEMBER]]
// CK1-DAG: br i1 true, label %[[LTYPE:[^,]+]], label %[[MEMBERCOM:[^,]+]]
// CK1-DAG: [[MEMBERCOM]]
// CK1-DAG: [[MEMBERCOMTYPE:%.+]] = add nuw i64 32, [[SHIPRESIZE]]
// CK1-DAG: br label %[[LTYPE]]
// CK1-DAG: [[LTYPE]]
// CK1-DAG: [[MEMBERTYPE:%.+]] = phi i64 [ 32, %[[MEMBER]] ], [ [[MEMBERCOMTYPE]], %[[MEMBERCOM]] ]
// CK1-DAG: [[TYPETF:%.+]] = and i64 [[TYPE]], 3
// CK1-DAG: [[ISALLOC:%.+]] = icmp eq i64 [[TYPETF]], 0
// CK1-DAG: br i1 [[ISALLOC]], label %[[ALLOC:[^,]+]], label %[[ALLOCELSE:[^,]+]]
// CK1-DAG: [[ALLOC]]
// CK1-DAG: [[ALLOCTYPE:%.+]] = and i64 [[MEMBERTYPE]], -4
// CK1-DAG: br label %[[TYEND:[^,]+]]
// CK1-DAG: [[ALLOCELSE]]
// CK1-DAG: [[ISTO:%.+]] = icmp eq i64 [[TYPETF]], 1
// CK1-DAG: br i1 [[ISTO]], label %[[TO:[^,]+]], label %[[TOELSE:[^,]+]]
// CK1-DAG: [[TO]]
// CK1-DAG: [[TOTYPE:%.+]] = and i64 [[MEMBERTYPE]], -3
// CK1-DAG: br label %[[TYEND]]
// CK1-DAG: [[TOELSE]]
// CK1-DAG: [[ISFROM:%.+]] = icmp eq i64 [[TYPETF]], 2
// CK1-DAG: br i1 [[ISFROM]], label %[[FROM:[^,]+]], label %[[TYEND]]
// CK1-DAG: [[FROM]]
// CK1-DAG: [[FROMTYPE:%.+]] = and i64 [[MEMBERTYPE]], -2
// CK1-DAG: br label %[[TYEND]]
// CK1-DAG: [[TYEND]]
// CK1-DAG: [[TYPE0:%.+]] = phi i64 [ [[ALLOCTYPE]], %[[ALLOC]] ], [ [[TOTYPE]], %[[TO]] ], [ [[FROMTYPE]], %[[FROM]] ], [ [[MEMBERTYPE]], %[[TOELSE]] ]
// CK1-64: call void @__tgt_push_mapper_component(i8* [[HANDLE]], i8* [[BPTRADDR0BC]], i8* [[PTRADDR0BC]], i64 [[CUSIZE]], i64 [[TYPE0]])
// CK1-DAG: [[BPTRADDR1BC:%.+]] = bitcast %class.C* [[OBJ]] to i8* // CK1-DAG: [[BPTRADDR1BC:%.+]] = bitcast %class.C* [[OBJ]] to i8*
// CK1-DAG: [[PTRADDR1BC:%.+]] = bitcast i32* [[ABEGIN]] to i8* // CK1-DAG: [[PTRADDR1BC:%.+]] = bitcast i32* [[ABEGIN]] to i8*
// CK1-DAG: br label %[[MEMBER:[^,]+]] // CK1-DAG: br label %[[MEMBER:[^,]+]]
// CK1-DAG: [[MEMBER]] // CK1-DAG: [[MEMBER]]
// CK1-DAG: br i1 false, label %[[LTYPE:[^,]+]], label %[[MEMBERCOM:[^,]+]] // CK1-DAG: br i1 true, label %[[LTYPE:[^,]+]], label %[[MEMBERCOM:[^,]+]]
// CK1-DAG: [[MEMBERCOM]] // CK1-DAG: [[MEMBERCOM]]
// 281474976710659 == 0x1,000,000,003 // CK1-DAG: [[MEMBERCOMTYPE:%.+]] = add nuw i64 35, [[SHIPRESIZE]]
// CK1-DAG: [[MEMBERCOMTYPE:%.+]] = add nuw i64 281474976710659, [[SHIPRESIZE]]
// CK1-DAG: br label %[[LTYPE]] // CK1-DAG: br label %[[LTYPE]]
// CK1-DAG: [[LTYPE]] // CK1-DAG: [[LTYPE]]
// CK1-DAG: [[MEMBERTYPE:%.+]] = phi i64 [ 281474976710659, %[[MEMBER]] ], [ [[MEMBERCOMTYPE]], %[[MEMBERCOM]] ] // CK1-DAG: [[MEMBERTYPE:%.+]] = phi i64 [ 35, %[[MEMBER]] ], [ [[MEMBERCOMTYPE]], %[[MEMBERCOM]] ]
// CK1-DAG: [[TYPETF:%.+]] = and i64 [[TYPE]], 3 // CK1-DAG: [[TYPETF:%.+]] = and i64 [[TYPE]], 3
// CK1-DAG: [[ISALLOC:%.+]] = icmp eq i64 [[TYPETF]], 0 // CK1-DAG: [[ISALLOC:%.+]] = icmp eq i64 [[TYPETF]], 0
// CK1-DAG: br i1 [[ISALLOC]], label %[[ALLOC:[^,]+]], label %[[ALLOCELSE:[^,]+]] // CK1-DAG: br i1 [[ISALLOC]], label %[[ALLOC:[^,]+]], label %[[ALLOCELSE:[^,]+]]
// CK1-DAG: [[ALLOC]] // CK1-DAG: [[ALLOC]]
// CK1-DAG: [[ALLOCTYPE:%.+]] = and i64 [[MEMBERTYPE]], -4 // CK1-DAG: [[ALLOCTYPE:%.+]] = and i64 [[MEMBERTYPE]], -4
// CK1-DAG: br label %[[TYEND:[^,]+]] // CK1-DAG: br label %[[TYEND:[^,]+]]
// CK1-DAG: [[ALLOCELSE]] // CK1-DAG: [[ALLOCELSE]]
// CK1-DAG: [[ISTO:%.+]] = icmp eq i64 [[TYPETF]], 1 // CK1-DAG: [[ISTO:%.+]] = icmp eq i64 [[TYPETF]], 1
▲ Show 20 Lines • Show All 96 Lines • ▼ Show 20 Lines
// CK2: [[LHEAD]] // CK2: [[LHEAD]]
// CK2: [[ISEMPTY:%.+]] = icmp eq %class.C** [[PTRBEGIN]], [[PTREND]] // CK2: [[ISEMPTY:%.+]] = icmp eq %class.C** [[PTRBEGIN]], [[PTREND]]
// CK2: br i1 [[ISEMPTY]], label %[[DONE:[^,]+]], label %[[LBODY:[^,]+]] // CK2: br i1 [[ISEMPTY]], label %[[DONE:[^,]+]], label %[[LBODY:[^,]+]]
// CK2: [[LBODY]] // CK2: [[LBODY]]
// CK2: [[PTR:%.+]] = phi %class.C** [ [[PTRBEGIN]], %[[LHEAD]] ], [ [[PTRNEXT:%.+]], %[[LCORRECT:[^,]+]] ] // CK2: [[PTR:%.+]] = phi %class.C** [ [[PTRBEGIN]], %[[LHEAD]] ], [ [[PTRNEXT:%.+]], %[[LCORRECT:[^,]+]] ]
// CK2: [[OBJ:%.+]] = load %class.C*, %class.C** [[PTR]] // CK2: [[OBJ:%.+]] = load %class.C*, %class.C** [[PTR]]
// CK2-DAG: [[BBEGIN:%.+]] = getelementptr inbounds %class.C, %class.C* [[OBJ]], i32 0, i32 1 // CK2-DAG: [[BBEGIN:%.+]] = getelementptr inbounds %class.C, %class.C* [[OBJ]], i32 0, i32 1
// CK2-DAG: [[BEND:%.+]] = getelementptr %class.B, %class.B* [[BBEGIN]], i32 1
// CK2-DAG: [[BBEGINV:%.+]] = bitcast %class.B* [[BBEGIN]] to i8*
// CK2-DAG: [[BENDV:%.+]] = bitcast %class.B* [[BEND]] to i8*
// CK2-DAG: [[BBEGINI:%.+]] = ptrtoint i8* [[BBEGINV]] to i64
// CK2-DAG: [[BENDI:%.+]] = ptrtoint i8* [[BENDV]] to i64
// CK2-DAG: [[BSIZE:%.+]] = sub i64 [[BENDI]], [[BBEGINI]]
// CK2-DAG: [[BUSIZE:%.+]] = sdiv exact i64 [[BSIZE]], ptrtoint (i8* getelementptr (i8, i8* null, i32 1) to i64)
// CK2-DAG: [[BPTRADDR0BC:%.+]] = bitcast %class.C* [[OBJ]] to i8*
// CK2-DAG: [[PTRADDR0BC:%.+]] = bitcast %class.B* [[BBEGIN]] to i8*
// CK2-DAG: [[PRESIZE:%.+]] = call i64 @__tgt_mapper_num_components(i8* [[HANDLE]]) // CK2-DAG: [[PRESIZE:%.+]] = call i64 @__tgt_mapper_num_components(i8* [[HANDLE]])
// CK2-DAG: [[SHIPRESIZE:%.+]] = shl i64 [[PRESIZE]], 48 // CK2-DAG: [[SHIPRESIZE:%.+]] = shl i64 [[PRESIZE]], 48
// CK2-DAG: br label %[[MEMBER:[^,]+]]
// CK2-DAG: [[MEMBER]]
// CK2-DAG: br i1 true, label %[[LTYPE:[^,]+]], label %[[MEMBERCOM:[^,]+]]
// CK2-DAG: [[MEMBERCOM]]
// CK2-DAG: [[MEMBERCOMTYPE:%.+]] = add nuw i64 32, [[SHIPRESIZE]]
// CK2-DAG: br label %[[LTYPE]]
// CK2-DAG: [[LTYPE]]
// CK2-DAG: [[MEMBERTYPE:%.+]] = phi i64 [ 32, %[[MEMBER]] ], [ [[MEMBERCOMTYPE]], %[[MEMBERCOM]] ]
// CK2-DAG: [[TYPETF:%.+]] = and i64 [[TYPE]], 3
// CK2-DAG: [[ISALLOC:%.+]] = icmp eq i64 [[TYPETF]], 0
// CK2-DAG: br i1 [[ISALLOC]], label %[[ALLOC:[^,]+]], label %[[ALLOCELSE:[^,]+]]
// CK2-DAG: [[ALLOC]]
// CK2-DAG: [[ALLOCTYPE:%.+]] = and i64 [[MEMBERTYPE]], -4
// CK2-DAG: br label %[[TYEND:[^,]+]]
// CK2-DAG: [[ALLOCELSE]]
// CK2-DAG: [[ISTO:%.+]] = icmp eq i64 [[TYPETF]], 1
// CK2-DAG: br i1 [[ISTO]], label %[[TO:[^,]+]], label %[[TOELSE:[^,]+]]
// CK2-DAG: [[TO]]
// CK2-DAG: [[TOTYPE:%.+]] = and i64 [[MEMBERTYPE]], -3
// CK2-DAG: br label %[[TYEND]]
// CK2-DAG: [[TOELSE]]
// CK2-DAG: [[ISFROM:%.+]] = icmp eq i64 [[TYPETF]], 2
// CK2-DAG: br i1 [[ISFROM]], label %[[FROM:[^,]+]], label %[[TYEND]]
// CK2-DAG: [[FROM]]
// CK2-DAG: [[FROMTYPE:%.+]] = and i64 [[MEMBERTYPE]], -2
// CK2-DAG: br label %[[TYEND]]
// CK2-DAG: [[TYEND]]
// CK2-DAG: [[TYPE0:%.+]] = phi i64 [ [[ALLOCTYPE]], %[[ALLOC]] ], [ [[TOTYPE]], %[[TO]] ], [ [[FROMTYPE]], %[[FROM]] ], [ [[MEMBERTYPE]], %[[TOELSE]] ]
// CK2-64: call void @__tgt_push_mapper_component(i8* [[HANDLE]], i8* [[BPTRADDR0BC]], i8* [[PTRADDR0BC]], i64 [[BUSIZE]], i64 [[TYPE0]])
// CK2-DAG: [[BPTRADDR1BC:%.+]] = bitcast %class.C* [[OBJ]] to i8* // CK2-DAG: [[BPTRADDR1BC:%.+]] = bitcast %class.C* [[OBJ]] to i8*
// CK2-DAG: [[PTRADDR1BC:%.+]] = bitcast %class.B* [[BBEGIN]] to i8* // CK2-DAG: [[PTRADDR1BC:%.+]] = bitcast %class.B* [[BBEGIN]] to i8*
// CK2-DAG: br label %[[MEMBER:[^,]+]] // CK2-DAG: br label %[[MEMBER:[^,]+]]
// CK2-DAG: [[MEMBER]] // CK2-DAG: [[MEMBER]]
// CK2-DAG: br i1 false, label %[[LTYPE:[^,]+]], label %[[MEMBERCOM:[^,]+]] // CK2-DAG: br i1 true, label %[[LTYPE:[^,]+]], label %[[MEMBERCOM:[^,]+]]
// CK2-DAG: [[MEMBERCOM]] // CK2-DAG: [[MEMBERCOM]]
// 281474976710659 == 0x1,000,000,003 // CK2-DAG: [[MEMBERCOMTYPE:%.+]] = add nuw i64 35, [[SHIPRESIZE]]
// CK2-DAG: [[MEMBERCOMTYPE:%.+]] = add nuw i64 281474976710659, [[SHIPRESIZE]]
// CK2-DAG: br label %[[LTYPE]] // CK2-DAG: br label %[[LTYPE]]
// CK2-DAG: [[LTYPE]] // CK2-DAG: [[LTYPE]]
// CK2-DAG: [[MEMBERTYPE:%.+]] = phi i64 [ 281474976710659, %[[MEMBER]] ], [ [[MEMBERCOMTYPE]], %[[MEMBERCOM]] ] // CK2-DAG: [[MEMBERTYPE:%.+]] = phi i64 [ 35, %[[MEMBER]] ], [ [[MEMBERCOMTYPE]], %[[MEMBERCOM]] ]
// CK2-DAG: [[TYPETF:%.+]] = and i64 [[TYPE]], 3 // CK2-DAG: [[TYPETF:%.+]] = and i64 [[TYPE]], 3
// CK2-DAG: [[ISALLOC:%.+]] = icmp eq i64 [[TYPETF]], 0 // CK2-DAG: [[ISALLOC:%.+]] = icmp eq i64 [[TYPETF]], 0
// CK2-DAG: br i1 [[ISALLOC]], label %[[ALLOC:[^,]+]], label %[[ALLOCELSE:[^,]+]] // CK2-DAG: br i1 [[ISALLOC]], label %[[ALLOC:[^,]+]], label %[[ALLOCELSE:[^,]+]]
// CK2-DAG: [[ALLOC]] // CK2-DAG: [[ALLOC]]
// CK2-DAG: [[ALLOCTYPE:%.+]] = and i64 [[MEMBERTYPE]], -4 // CK2-DAG: [[ALLOCTYPE:%.+]] = and i64 [[MEMBERTYPE]], -4
// CK2-DAG: br label %[[TYEND:[^,]+]] // CK2-DAG: br label %[[TYEND:[^,]+]]
// CK2-DAG: [[ALLOCELSE]] // CK2-DAG: [[ALLOCELSE]]
// CK2-DAG: [[ISTO:%.+]] = icmp eq i64 [[TYPETF]], 1 // CK2-DAG: [[ISTO:%.+]] = icmp eq i64 [[TYPETF]], 1
▲ Show 20 Lines • Show All 46 Lines • ▼ Show 20 Lines
// RUN: %clang_cc1 -DCK3 -verify -fopenmp-simd -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm -femit-all-decls -disable-llvm-passes %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s // RUN: %clang_cc1 -DCK3 -verify -fopenmp-simd -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm -femit-all-decls -disable-llvm-passes %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s
// RUN: %clang_cc1 -DCK3 -fopenmp-simd -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -femit-all-decls -disable-llvm-passes -o %t %s // RUN: %clang_cc1 -DCK3 -fopenmp-simd -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -femit-all-decls -disable-llvm-passes -o %t %s
// RUN: %clang_cc1 -DCK3 -fopenmp-simd -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -femit-all-decls -disable-llvm-passes -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s // RUN: %clang_cc1 -DCK3 -fopenmp-simd -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -femit-all-decls -disable-llvm-passes -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s
#ifdef CK3 #ifdef CK3
// map of array sections and nested components. // map of array sections and nested components.
// CK3-LABEL: @.__omp_offloading_{{.*}}foo{{.*}}.region_id = weak constant i8 0 // CK3-LABEL: @.__omp_offloading_{{.*}}foo{{.*}}.region_id = weak constant i8 0
// CK3: [[TYPES:@.+]] = {{.+}}constant [3 x i64] [i64 32, i64 281474976710659, i64 35] // CK3-DAG: [[SIZES:@.+]] = {{.+}}constant [2 x i64] [i64 {{8|16}}, i64 {{80|160}}]
// CK3-DAG: [[TYPES:@.+]] = {{.+}}constant [2 x i64] [i64 35, i64 35]
class C { class C {
public: public:
int a; int a;
double *b; double *b;
}; };
class B { class B {
public: public:
C c; C c;
}; };
#pragma omp declare mapper(id: C s) map(s.a, s.b[0:2]) #pragma omp declare mapper(id: C s) map(s.a, s.b[0:2])
// CK3: define {{.*}}void [[MPRFUNC:@[.]omp_mapper[.].*C[.]id]](i8*{{.*}}, i8*{{.*}}, i8*{{.*}}, i64{{.*}}, i64{{.*}}) // CK3: define {{.*}}void [[MPRFUNC:@[.]omp_mapper[.].*C[.]id]](i8*{{.*}}, i8*{{.*}}, i8*{{.*}}, i64{{.*}}, i64{{.*}})
// CK3-LABEL: define {{.*}}void @{{.*}}foo{{.*}} // CK3-LABEL: define {{.*}}void @{{.*}}foo{{.*}}
void foo(int a){ void foo(int a){
// CK3-DAG: [[CVAL:%.+]] = alloca [10 x %class.C] // CK3-DAG: [[CVAL:%.+]] = alloca [10 x %class.C]
// CK3-DAG: [[BVAL:%.+]] = alloca %class.B // CK3-DAG: [[BVAL:%.+]] = alloca %class.B
C c[10]; C c[10];
B b; B b;
// CK3-DAG: [[BC:%.+]] = getelementptr inbounds %class.B, %class.B* [[BVAL]], i32 0, i32 0 // CK3-DAG: [[BC:%.+]] = getelementptr inbounds %class.B, %class.B* [[BVAL]], i32 0, i32 0
// CK3-DAG: [[BCEND:%.+]] = getelementptr %class.C, %class.C* [[BC]], i32 1
// CK3-DAG: [[BCC:%.+]] = bitcast %class.C* [[BC]] to i8*
// CK3-DAG: [[BCENDC:%.+]] = bitcast %class.C* [[BCEND]] to i8*
// CK3-DAG: [[BCI:%.+]] = ptrtoint i8* [[BCC]] to i64
// CK3-DAG: [[BCENDI:%.+]] = ptrtoint i8* [[BCENDC]] to i64
// CK3-DAG: [[BSIZE:%.+]] = sub i64 [[BCENDI]], [[BCI]]
// CK3-DAG: [[BSIZED:%.+]] = sdiv exact i64 [[BSIZE]], ptrtoint (i8* getelementptr (i8, i8* null, i32 1) to i64)
// CK3-DAG: call i32 @__tgt_target_mapper(i64 {{.+}}, i8* {{.+}}, i32 3, i8** [[BPGEP:%[0-9]+]], i8** [[PGEP:%[0-9]+]], i64* [[SGEP:%[^,]+]], {{.+}}[[TYPES]]{{.+}}, i8** [[MPRGEP:%.+]]) // CK3-DAG: call i32 @__tgt_target_mapper(i64 {{.+}}, i8* {{.+}}, i32 2, i8** [[BPGEP:%[0-9]+]], i8** [[PGEP:%[0-9]+]], {{.+}}[[SIZES]]{{.+}}, {{.+}}[[TYPES]]{{.+}}, i8** [[MPRGEP:%.+]])
// CK3-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BPS:%[^,]+]], i32 0, i32 0 // CK3-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BPS:%[^,]+]], i32 0, i32 0
// CK3-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[PS:%[^,]+]], i32 0, i32 0 // CK3-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[PS:%[^,]+]], i32 0, i32 0
// CK3-DAG: [[SGEP]] = getelementptr inbounds {{.+}}[[SIZES:%[^,]+]], i32 0, i32 0 // CK3-DAG: [[MPRGEP]] = bitcast [2 x i8*]* [[MPR:%[^,]+]] to i8**
// CK3-DAG: [[MPRGEP]] = bitcast [3 x i8*]* [[MPR:%[^,]+]] to i8**
// CK3-DAG: [[BP1:%.+]] = getelementptr inbounds {{.+}}[[BPS]], i32 0, i32 0 // CK3-DAG: [[BP1:%.+]] = getelementptr inbounds {{.+}}[[BPS]], i32 0, i32 0
// CK3-DAG: [[P1:%.+]] = getelementptr inbounds {{.+}}[[PS]], i32 0, i32 0 // CK3-DAG: [[P1:%.+]] = getelementptr inbounds {{.+}}[[PS]], i32 0, i32 0
// CK3-DAG: [[S1:%.+]] = getelementptr inbounds {{.+}}[[SIZES]], i32 0, i32 0
// CK3-DAG: [[MPR1:%.+]] = getelementptr inbounds {{.+}}[[MPR]], i{{64|32}} 0, i{{64|32}} 0 // CK3-DAG: [[MPR1:%.+]] = getelementptr inbounds {{.+}}[[MPR]], i{{64|32}} 0, i{{64|32}} 0
// CK3-DAG: [[CBP1:%.+]] = bitcast i8** [[BP1]] to %class.B** // CK3-DAG: [[CBP1:%.+]] = bitcast i8** [[BP1]] to %class.B**
// CK3-DAG: [[CP1:%.+]] = bitcast i8** [[P1]] to %class.C** // CK3-DAG: [[CP1:%.+]] = bitcast i8** [[P1]] to %class.C**
// CK3-DAG: store %class.B* [[BVAL]], %class.B** [[CBP1]] // CK3-DAG: store %class.B* [[BVAL]], %class.B** [[CBP1]]
// CK3-DAG: store %class.C* [[BC]], %class.C** [[CP1]] // CK3-DAG: store %class.C* [[BC]], %class.C** [[CP1]]
// CK3-DAG: store i64 [[BSIZED]], i64* [[S1]] // CK3-DAG: store i8* bitcast (void (i8*, i8*, i8*, i64, i64)* [[MPRFUNC]] to i8*), i8** [[MPR1]]
// CK3-DAG: store i8* null, i8** [[MPR1]]
// CK3-DAG: [[BP2:%.+]] = getelementptr inbounds {{.+}}[[BPS]], i32 0, i32 1 // CK3-DAG: [[BP2:%.+]] = getelementptr inbounds {{.+}}[[BPS]], i32 0, i32 1
// CK3-DAG: [[P2:%.+]] = getelementptr inbounds {{.+}}[[PS]], i32 0, i32 1 // CK3-DAG: [[P2:%.+]] = getelementptr inbounds {{.+}}[[PS]], i32 0, i32 1
// CK3-DAG: [[S2:%.+]] = getelementptr inbounds {{.+}}[[SIZES]], i32 0, i32 1
// CK3-DAG: [[MPR2:%.+]] = getelementptr inbounds {{.+}}[[MPR]], i{{64|32}} 0, i{{64|32}} 1 // CK3-DAG: [[MPR2:%.+]] = getelementptr inbounds {{.+}}[[MPR]], i{{64|32}} 0, i{{64|32}} 1
// CK3-DAG: [[CBP2:%.+]] = bitcast i8** [[BP2]] to %class.B** // CK3-DAG: [[CBP2:%.+]] = bitcast i8** [[BP2]] to [10 x %class.C]**
// CK3-DAG: [[CP2:%.+]] = bitcast i8** [[P2]] to %class.C** // CK3-DAG: [[CP2:%.+]] = bitcast i8** [[P2]] to %class.C**
// CK3-DAG: store %class.B* [[BVAL]], %class.B** [[CBP2]] // CK3-DAG: store [10 x %class.C]* [[CVAL]], [10 x %class.C]** [[CBP2]]
// CK3-DAG: store %class.C* [[BC]], %class.C** [[CP2]]
// CK3-64-DAG: store i64 16, i64* [[S2]]
// CK3-32-DAG: store i64 8, i64* [[S2]]
// CK3-DAG: store i8* bitcast (void (i8*, i8*, i8*, i64, i64)* [[MPRFUNC]] to i8*), i8** [[MPR2]]
// CK3-DAG: [[BP3:%.+]] = getelementptr inbounds {{.+}}[[BPS]], i32 0, i32 2
// CK3-DAG: [[P3:%.+]] = getelementptr inbounds {{.+}}[[PS]], i32 0, i32 2
// CK3-DAG: [[S3:%.+]] = getelementptr inbounds {{.+}}[[SIZES]], i32 0, i32 2
// CK3-DAG: [[MPR3:%.+]] = getelementptr inbounds {{.+}}[[MPR]], i{{64|32}} 0, i{{64|32}} 2
// CK3-DAG: [[CBP3:%.+]] = bitcast i8** [[BP3]] to [10 x %class.C]**
// CK3-DAG: [[CP3:%.+]] = bitcast i8** [[P3]] to %class.C**
// CK3-DAG: store [10 x %class.C]* [[CVAL]], [10 x %class.C]** [[CBP3]]
// CK3-DAG: [[CVALGEP:%.+]] = getelementptr inbounds {{.+}}[[CVAL]], i{{64|32}} 0, i{{64|32}} 0 // CK3-DAG: [[CVALGEP:%.+]] = getelementptr inbounds {{.+}}[[CVAL]], i{{64|32}} 0, i{{64|32}} 0
// CK3-DAG: store %class.C* [[CVALGEP]], %class.C** [[CP3]] // CK3-DAG: store %class.C* [[CVALGEP]], %class.C** [[CP2]]
// CK3-64-DAG: store i64 160, i64* [[S3]] // CK3-DAG: store i8* bitcast (void (i8*, i8*, i8*, i64, i64)* [[MPRFUNC]] to i8*), i8** [[MPR2]]
// CK3-32-DAG: store i64 80, i64* [[S3]]
// CK3-DAG: store i8* bitcast (void (i8*, i8*, i8*, i64, i64)* [[MPRFUNC]] to i8*), i8** [[MPR3]]
// CK3: call void [[KERNEL:@.+]](%class.B* [[BVAL]], [10 x %class.C]* [[CVAL]]) // CK3: call void [[KERNEL:@.+]](%class.B* [[BVAL]], [10 x %class.C]* [[CVAL]])
#pragma omp target map(mapper(id),tofrom: c[0:10], b.c) #pragma omp target map(mapper(id),tofrom: c[0:10], b.c)
for (int i = 0; i < 10; i++) { for (int i = 0; i < 10; i++) {
b.c.a += ++c[i].a; b.c.a += ++c[i].a;
} }
} }
▲ Show 20 Lines • Show All 243 Lines • Show Last 20 Lines

clang/test/OpenMP/target_codegen.cpp

Show First 20 Lines • Show All 75 Lines • ▼ Show 20 Lines
// CHECK-DAG: [[MAPT2:@.+]] = private unnamed_addr constant [1 x i64] [i64 800] // CHECK-DAG: [[MAPT2:@.+]] = private unnamed_addr constant [1 x i64] [i64 800]
// CHECK-DAG: [[SIZET3:@.+]] = private unnamed_addr constant [2 x i64] [i64 4, i64 2] // CHECK-DAG: [[SIZET3:@.+]] = private unnamed_addr constant [2 x i64] [i64 4, i64 2]
// CHECK-DAG: [[MAPT3:@.+]] = private unnamed_addr constant [2 x i64] [i64 800, i64 800] // CHECK-DAG: [[MAPT3:@.+]] = private unnamed_addr constant [2 x i64] [i64 800, i64 800]
// CHECK-DAG: [[MAPT4:@.+]] = private unnamed_addr constant [9 x i64] [i64 800, i64 547, i64 800, i64 547, i64 547, i64 800, i64 800, i64 547, i64 547] // CHECK-DAG: [[MAPT4:@.+]] = private unnamed_addr constant [9 x i64] [i64 800, i64 547, i64 800, i64 547, i64 547, i64 800, i64 800, i64 547, i64 547]
// CHECK-DAG: [[SIZET5:@.+]] = private unnamed_addr constant [3 x i64] [i64 4, i64 2, i64 40] // CHECK-DAG: [[SIZET5:@.+]] = private unnamed_addr constant [3 x i64] [i64 4, i64 2, i64 40]
// CHECK-DAG: [[MAPT5:@.+]] = private unnamed_addr constant [3 x i64] [i64 800, i64 800, i64 547] // CHECK-DAG: [[MAPT5:@.+]] = private unnamed_addr constant [3 x i64] [i64 800, i64 800, i64 547]
// CHECK-DAG: [[SIZET6:@.+]] = private unnamed_addr constant [4 x i64] [i64 4, i64 2, i64 1, i64 40] // CHECK-DAG: [[SIZET6:@.+]] = private unnamed_addr constant [4 x i64] [i64 4, i64 2, i64 1, i64 40]
// CHECK-DAG: [[MAPT6:@.+]] = private unnamed_addr constant [4 x i64] [i64 800, i64 800, i64 800, i64 547] // CHECK-DAG: [[MAPT6:@.+]] = private unnamed_addr constant [4 x i64] [i64 800, i64 800, i64 800, i64 547]
// CHECK-DAG: [[MAPT7:@.+]] = private unnamed_addr constant [6 x i64] [i64 32, i64 281474976711171, i64 800, i64 800, i64 800, i64 547] // CHECK-DAG: [[MAPT7:@.+]] = private unnamed_addr constant [5 x i64] [i64 547, i64 800, i64 800, i64 800, i64 547]
// CHECK-DAG: [[SIZET9:@.+]] = private unnamed_addr constant [1 x i64] [i64 12] // CHECK-DAG: [[SIZET9:@.+]] = private unnamed_addr constant [1 x i64] [i64 12]
// CHECK-DAG: [[MAPT10:@.+]] = private unnamed_addr constant [1 x i64] [i64 35] // CHECK-DAG: [[MAPT10:@.+]] = private unnamed_addr constant [1 x i64] [i64 35]
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
▲ Show 20 Lines • Show All 481 Lines • ▼ Show 20 Lines
// CHECK: br i1 [[IF]], label %[[TRY:[^,]+]], label %[[IFELSE:[^,]+]] // CHECK: br i1 [[IF]], label %[[TRY:[^,]+]], label %[[IFELSE:[^,]+]]
// CHECK: [[TRY]] // CHECK: [[TRY]]
// We capture 2 VLA sizes in this target region // We capture 2 VLA sizes in this target region
// CHECK: [[CELEMSIZE2:%.+]] = mul nuw i[[SZ]] 2, [[VLA0:%.+]] // CHECK: [[CELEMSIZE2:%.+]] = mul nuw i[[SZ]] 2, [[VLA0:%.+]]
// CHECK-64: [[CSIZE:%.+]] = mul nuw i64 [[CELEMSIZE2]], 2 // CHECK-64: [[CSIZE:%.+]] = mul nuw i64 [[CELEMSIZE2]], 2
// CHECK-32: [[CSZSIZE:%.+]] = mul nuw i32 [[CELEMSIZE2]], 2 // CHECK-32: [[CSZSIZE:%.+]] = mul nuw i32 [[CELEMSIZE2]], 2
// CHECK-32: [[CSIZE:%.+]] = sext i32 [[CSZSIZE]] to i64 // CHECK-32: [[CSIZE:%.+]] = sext i32 [[CSZSIZE]] to i64
// CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_mapper(i64 -1, i8* @{{[^,]+}}, i32 6, i8** [[BPR:%[^,]+]], i8** [[PR:%[^,]+]], i64* [[SR:%[^,]+]], i64* getelementptr inbounds ([6 x i64], [6 x i64]* [[MAPT7]], i32 0, i32 0), i8** null) // CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_mapper(i64 -1, i8* @{{[^,]+}}, i32 5, i8** [[BPR:%[^,]+]], i8** [[PR:%[^,]+]], i64* [[SR:%[^,]+]], i64* getelementptr inbounds ([5 x i64], [5 x i64]* [[MAPT7]], i32 0, i32 0), i8** null)
// CHECK-DAG: [[BPR]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP:%.+]], i32 0, i32 0 // CHECK-DAG: [[BPR]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP:%.+]], i32 0, i32 0
// CHECK-DAG: [[PR]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P:%.+]], i32 0, i32 0 // CHECK-DAG: [[PR]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P:%.+]], i32 0, i32 0
// CHECK-DAG: [[SR]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S:%.+]], i32 0, i32 0 // CHECK-DAG: [[SR]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S:%.+]], i32 0, i32 0
// CHECK-DAG: [[SADDR0:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 0, i32 [[IDX0:0]] // CHECK-DAG: [[SADDR0:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 0, i32 [[IDX0:0]]
// CHECK-DAG: [[BPADDR0:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 0, i32 [[IDX0]] // CHECK-DAG: [[BPADDR0:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 0, i32 [[IDX0]]
// CHECK-DAG: [[PADDR0:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 0, i32 [[IDX0]] // CHECK-DAG: [[PADDR0:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 0, i32 [[IDX0]]
// CHECK-DAG: [[SADDR1:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 0, i32 [[IDX1:1]] // CHECK-DAG: [[SADDR1:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 0, i32 [[IDX1:1]]
// CHECK-DAG: [[BPADDR1:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 0, i32 [[IDX1]] // CHECK-DAG: [[BPADDR1:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 0, i32 [[IDX1]]
// CHECK-DAG: [[PADDR1:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 0, i32 [[IDX1]] // CHECK-DAG: [[PADDR1:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 0, i32 [[IDX1]]
// CHECK-DAG: [[SADDR2:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 0, i32 [[IDX2:2]] // CHECK-DAG: [[SADDR2:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 0, i32 [[IDX2:2]]
// CHECK-DAG: [[BPADDR2:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 0, i32 [[IDX2]] // CHECK-DAG: [[BPADDR2:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 0, i32 [[IDX2]]
// CHECK-DAG: [[PADDR2:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 0, i32 [[IDX2]] // CHECK-DAG: [[PADDR2:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 0, i32 [[IDX2]]
// CHECK-DAG: [[SADDR3:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 0, i32 [[IDX3:3]] // CHECK-DAG: [[SADDR3:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 0, i32 [[IDX3:3]]
// CHECK-DAG: [[BPADDR3:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 0, i32 [[IDX3]] // CHECK-DAG: [[BPADDR3:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 0, i32 [[IDX3]]
// CHECK-DAG: [[PADDR3:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 0, i32 [[IDX3]] // CHECK-DAG: [[PADDR3:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 0, i32 [[IDX3]]
// CHECK-DAG: [[SADDR4:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 0, i32 [[IDX4:4]] // CHECK-DAG: [[SADDR4:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 0, i32 [[IDX4:4]]
// CHECK-DAG: [[BPADDR4:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 0, i32 [[IDX4]] // CHECK-DAG: [[BPADDR4:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 0, i32 [[IDX4]]
// CHECK-DAG: [[PADDR4:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 0, i32 [[IDX4]] // CHECK-DAG: [[PADDR4:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 0, i32 [[IDX4]]
// CHECK-DAG: [[SADDR5:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 0, i32 [[IDX5:5]]
// CHECK-DAG: [[BPADDR5:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 0, i32 [[IDX5]]
// CHECK-DAG: [[PADDR5:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 0, i32 [[IDX5]]
// The names below are not necessarily consistent with the names used for the // The names below are not necessarily consistent with the names used for the
// addresses above as some are repeated. // addresses above as some are repeated.
// CHECK-DAG: [[CBPADDR5:%.+]] = bitcast i8** [[BPADDR5]] to i16** // CHECK-DAG: [[CBPADDR4:%.+]] = bitcast i8** [[BPADDR4]] to i16**
// CHECK-DAG: [[CPADDR5:%.+]] = bitcast i8** [[PADDR5]] to i16** // CHECK-DAG: [[CPADDR4:%.+]] = bitcast i8** [[PADDR4]] to i16**
// CHECK-DAG: store i16* %{{.+}}, i16** [[CBPADDR5]] // CHECK-DAG: store i16* %{{.+}}, i16** [[CBPADDR4]]
// CHECK-DAG: store i16* %{{.+}}, i16** [[CPADDR5]] // CHECK-DAG: store i16* %{{.+}}, i16** [[CPADDR4]]
// CHECK-DAG: store i64 [[CSIZE]], i64* [[SADDR5]] // CHECK-DAG: store i64 [[CSIZE]], i64* [[SADDR4]]
// CHECK-DAG: [[CBPADDR4:%.+]] = bitcast i8** [[BPADDR4]] to i[[SZ]]*
// CHECK-DAG: [[CPADDR4:%.+]] = bitcast i8** [[PADDR4]] to i[[SZ]]*
// CHECK-DAG: store i[[SZ]] [[VLA0]], i[[SZ]]* [[CBPADDR4]]
// CHECK-DAG: store i[[SZ]] [[VLA0]], i[[SZ]]* [[CPADDR4]]
// CHECK-DAG: store i64 {{4|8}}, i64* [[SADDR4]]
// CHECK-DAG: [[CBPADDR3:%.+]] = bitcast i8** [[BPADDR3]] to i[[SZ]]* // CHECK-DAG: [[CBPADDR3:%.+]] = bitcast i8** [[BPADDR3]] to i[[SZ]]*
// CHECK-DAG: [[CPADDR3:%.+]] = bitcast i8** [[PADDR3]] to i[[SZ]]* // CHECK-DAG: [[CPADDR3:%.+]] = bitcast i8** [[PADDR3]] to i[[SZ]]*
// CHECK-DAG: store i[[SZ]] 2, i[[SZ]]* [[CBPADDR3]] // CHECK-DAG: store i[[SZ]] [[VLA0]], i[[SZ]]* [[CBPADDR3]]
// CHECK-DAG: store i[[SZ]] 2, i[[SZ]]* [[CPADDR3]] // CHECK-DAG: store i[[SZ]] [[VLA0]], i[[SZ]]* [[CPADDR3]]
// CHECK-DAG: store i64 {{4|8}}, i64* [[SADDR3]] // CHECK-DAG: store i64 {{4|8}}, i64* [[SADDR3]]
// CHECK-DAG: [[CBPADDR2:%.+]] = bitcast i8** [[BPADDR2]] to i[[SZ]]* // CHECK-DAG: [[CBPADDR2:%.+]] = bitcast i8** [[BPADDR2]] to i[[SZ]]*
// CHECK-DAG: [[CPADDR2:%.+]] = bitcast i8** [[PADDR2]] to i[[SZ]]* // CHECK-DAG: [[CPADDR2:%.+]] = bitcast i8** [[PADDR2]] to i[[SZ]]*
// CHECK-DAG: store i[[SZ]] [[B_CVAL]], i[[SZ]]* [[CBPADDR2]] // CHECK-DAG: store i[[SZ]] 2, i[[SZ]]* [[CBPADDR2]]
// CHECK-DAG: store i[[SZ]] [[B_CVAL]], i[[SZ]]* [[CPADDR2]] // CHECK-DAG: store i[[SZ]] 2, i[[SZ]]* [[CPADDR2]]
// CHECK-DAG: store i64 4, i64* [[SADDR2]] // CHECK-DAG: store i64 {{4|8}}, i64* [[SADDR2]]
// CHECK-DAG: [[CBPADDR1:%.+]] = bitcast i8** [[BPADDR1]] to i[[SZ]]*
// CHECK-DAG: [[CPADDR1:%.+]] = bitcast i8** [[PADDR1]] to i[[SZ]]*
// CHECK-DAG: store i[[SZ]] [[B_CVAL]], i[[SZ]]* [[CBPADDR1]]
// CHECK-DAG: store i[[SZ]] [[B_CVAL]], i[[SZ]]* [[CPADDR1]]
// CHECK-DAG: store i64 4, i64* [[SADDR1]]
// CHECK-DAG: [[CBPADDR0:%.+]] = bitcast i8** [[BPADDR0]] to [[S1]]** // CHECK-DAG: [[CBPADDR0:%.+]] = bitcast i8** [[BPADDR0]] to [[S1]]**
// CHECK-DAG: [[CPADDR0:%.+]] = bitcast i8** [[PADDR0]] to double** // CHECK-DAG: [[CPADDR0:%.+]] = bitcast i8** [[PADDR0]] to double**
// CHECK-DAG: store [[S1]]* [[THIS:%.+]], [[S1]]** [[CBPADDR0]] // CHECK-DAG: store [[S1]]* [[THIS:%.+]], [[S1]]** [[CBPADDR0]]
// CHECK-DAG: store double* [[A:%.+]], double** [[CPADDR0]] // CHECK-DAG: store double* [[A:%.+]], double** [[CPADDR0]]
// CHECK-DAG: store i64 %{{.+}}, i64* [[SADDR0]] // CHECK-DAG: store i64 8, i64* [[SADDR0]]
// CHECK-DAG: [[CBPADDR1:%.+]] = bitcast i8** [[BPADDR1]] to [[S1]]**
// CHECK-DAG: [[CPADDR1:%.+]] = bitcast i8** [[PADDR1]] to double**
// CHECK-DAG: store [[S1]]* [[THIS]], [[S1]]** [[CBPADDR1]]
// CHECK-DAG: store double* [[A]], double** [[CPADDR1]]
// CHECK-DAG: store i64 8, i64* [[SADDR1]]
// CHECK: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0 // CHECK: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0
// CHECK-NEXT: br i1 [[ERROR]], label %[[FAIL:.+]], label %[[END:[^,]+]] // CHECK-NEXT: br i1 [[ERROR]], label %[[FAIL:.+]], label %[[END:[^,]+]]
// CHECK: [[FAIL]] // CHECK: [[FAIL]]
// CHECK: call void [[HVT7:@.+]]({{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}) // CHECK: call void [[HVT7:@.+]]({{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}})
// CHECK-NEXT: br label %[[END]] // CHECK-NEXT: br label %[[END]]
// CHECK: [[END]] // CHECK: [[END]]
// CHECK-NEXT: br label %[[IFEND:.+]] // CHECK-NEXT: br label %[[IFEND:.+]]
▲ Show 20 Lines • Show All 259 Lines • Show Last 20 Lines

clang/test/OpenMP/target_data_member_codegen.cpp

Show All 19 Lines
// RUN: %clang_cc1 -fopenmp-simd -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s // RUN: %clang_cc1 -fopenmp-simd -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s
// SIMD-ONLY0-NOT: {{__kmpc|__tgt}} // SIMD-ONLY0-NOT: {{__kmpc|__tgt}}
// expected-no-diagnostics // expected-no-diagnostics
#ifndef HEADER #ifndef HEADER
#define HEADER #define HEADER
// CHECK: [[MAPTYPES:@.+]] = private unnamed_addr constant [1 x i64] zeroinitializer // CHECK-DAG: [[SIZES:@.+]] = private unnamed_addr constant [1 x i64] [i64 {{8|4}}]
// CHECK-DAG: [[MAPTYPES:@.+]] = private unnamed_addr constant [1 x i64] [i64 3]
struct S { struct S {
double *p; double *p;
}; };
// CHECK-LABEL: @main // CHECK-LABEL: @main
int main() { int main() {
// CHECK: call i32 @__tgt_target_mapper(i64 -1, i8* @.{{.+}}, i32 1, i8** %{{.+}}, i8** %{{.+}}, i64* %{{.+}}, i64* getelementptr inbounds ([1 x i64], [1 x i64]* [[MAPTYPES]], i32 0, i32 0), i8** null) // CHECK: call i32 @__tgt_target_mapper(i64 -1, i8* @.{{.+}}, i32 1, i8** %{{.+}}, i8** %{{.+}}, i64* getelementptr inbounds ([1 x i64], [1 x i64]* [[SIZES]], i32 0, i32 0), i64* getelementptr inbounds ([1 x i64], [1 x i64]* [[MAPTYPES]], i32 0, i32 0), i8** null)
S s; S s;
#pragma omp target map(s.p) #pragma omp target map(s.p)
{ } { }
} }
#endif //HEADER #endif //HEADER

clang/test/OpenMP/target_firstprivate_codegen.cpp

Show First 20 Lines • Show All 52 Lines • ▼ Show 20 Lines
// TCHECK-DAG: [[S1:%.+]] = type { double } // TCHECK-DAG: [[S1:%.+]] = type { double }
// CHECK-DAG: [[FP_E:@__omp_offloading_firstprivate_.+_e_l76]] = internal global [[TTII]] zeroinitializer // CHECK-DAG: [[FP_E:@__omp_offloading_firstprivate_.+_e_l76]] = internal global [[TTII]] zeroinitializer
// CHECK-DAG: [[SIZET:@.+]] = private unnamed_addr constant [2 x i{{32|64}}] [i[[SZ:32|64]] 4, i{{64|32}} {{8|4}}] // CHECK-DAG: [[SIZET:@.+]] = private unnamed_addr constant [2 x i{{32|64}}] [i[[SZ:32|64]] 4, i{{64|32}} {{8|4}}]
// CHECK-DAG: [[MAPT:@.+]] = private unnamed_addr constant [2 x i64] [i64 288, i64 49] // CHECK-DAG: [[MAPT:@.+]] = private unnamed_addr constant [2 x i64] [i64 288, i64 49]
// CHECK-DAG: [[MAPT2:@.+]] = private unnamed_addr constant [9 x i64] [i64 288, i64 161, i64 800, i64 161, i64 161, i64 800, i64 800, i64 161, i64 161] // CHECK-DAG: [[MAPT2:@.+]] = private unnamed_addr constant [9 x i64] [i64 288, i64 161, i64 800, i64 161, i64 161, i64 800, i64 800, i64 161, i64 161]
// CHECK-DAG: [[SIZET3:@.+]] = private unnamed_addr constant [2 x i{{32|64}}] [i{{32|64}} 0, i{{32|64}} 8] // CHECK-DAG: [[SIZET3:@.+]] = private unnamed_addr constant [2 x i{{32|64}}] [i{{32|64}} 0, i{{32|64}} 8]
// CHECK-DAG: [[MAPT3:@.+]] = private unnamed_addr constant [2 x i64] [i64 32, i64 37] // CHECK-DAG: [[MAPT3:@.+]] = private unnamed_addr constant [2 x i64] [i64 32, i64 37]
// CHECK-DAG: [[MAPT4:@.+]] = private unnamed_addr constant [6 x i64] [i64 32, i64 281474976711171, i64 288, i64 800, i64 800, i64 161] // CHECK-DAG: [[MAPT4:@.+]] = private unnamed_addr constant [5 x i64] [i64 547, i64 288, i64 800, i64 800, i64 161]
// CHECK-DAG: [[SIZET5:@.+]] = private unnamed_addr constant [3 x i{{32|64}}] [i[[SZ]] 4, i[[SZ]] 1, i[[SZ]] 40] // CHECK-DAG: [[SIZET5:@.+]] = private unnamed_addr constant [3 x i{{32|64}}] [i[[SZ]] 4, i[[SZ]] 1, i[[SZ]] 40]
// CHECK-DAG: [[MAPT5:@.+]] = private unnamed_addr constant [3 x i64] [i64 288, i64 288, i64 161] // CHECK-DAG: [[MAPT5:@.+]] = private unnamed_addr constant [3 x i64] [i64 288, i64 288, i64 161]
// CHECK-DAG: [[SIZET6:@.+]] = private unnamed_addr constant [2 x i{{32|64}}] [i[[SZ]] 4, i[[SZ]] 40] // CHECK-DAG: [[SIZET6:@.+]] = private unnamed_addr constant [2 x i{{32|64}}] [i[[SZ]] 4, i[[SZ]] 40]
// CHECK-DAG: [[MAPT6:@.+]] = private unnamed_addr constant [2 x i64] [i64 288, i64 161] // CHECK-DAG: [[MAPT6:@.+]] = private unnamed_addr constant [2 x i64] [i64 288, i64 161]
// CHECK: define {{.*}}[[FOO:@.+]]( // CHECK: define {{.*}}[[FOO:@.+]](
int foo(int n, double *ptr) { int foo(int n, double *ptr) {
int a = 0; int a = 0;
▲ Show 20 Lines • Show All 363 Lines • ▼ Show 20 Lines#pragma omp target firstprivate(b, c)
c[1][1] = ++a; c[1][1] = ++a;
} }
return c[1][1] + (int)b; return c[1][1] + (int)b;
} }
// on the host side, we first generate r1, then the static function and the template above // on the host side, we first generate r1, then the static function and the template above
// CHECK: define{{.+}} i32 {{.+}}([[S1]]* {{.+}}, i{{[0-9]+}} {{.+}}) // CHECK: define{{.+}} i32 {{.+}}([[S1]]* {{.+}}, i{{[0-9]+}} {{.+}})
// CHECK: [[BASE_PTRS4:%.+]] = alloca [6 x i8*], // CHECK: [[BASE_PTRS4:%.+]] = alloca [5 x i8*],
// CHECK: [[PTRS4:%.+]] = alloca [6 x i8*], // CHECK: [[PTRS4:%.+]] = alloca [5 x i8*],
// CHECK: [[SIZET4:%.+]] = alloca [6 x i{{[0-9]+}}], // CHECK: [[SIZET4:%.+]] = alloca [5 x i{{[0-9]+}}],
// map(this: this ptr is implicitly captured (not firstprivate matter) // map(this: this ptr is implicitly captured (not firstprivate matter)
// CHECK: [[BP0:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BASE_PTRS4]], i{{[0-9]+}} 0, i{{[0-9]+}} 0 // CHECK: [[BP0:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BASE_PTRS4]], i{{[0-9]+}} 0, i{{[0-9]+}} 0
// CHECK: [[CBP0:%.+]] = bitcast i8** [[BP0]] to %struct.S1** // CHECK: [[CBP0:%.+]] = bitcast i8** [[BP0]] to %struct.S1**
// CHECK: store %struct.S1* [[THIS:%.+]], %struct.S1** [[CBP0]], // CHECK: store %struct.S1* [[THIS:%.+]], %struct.S1** [[CBP0]],
// CHECK: [[P0:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[PTRS4]], i{{[0-9]+}} 0, i{{[0-9]+}} 0 // CHECK: [[P0:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[PTRS4]], i{{[0-9]+}} 0, i{{[0-9]+}} 0
// CHECK: [[CP0:%.+]] = bitcast i8** [[P0]] to double** // CHECK: [[CP0:%.+]] = bitcast i8** [[P0]] to double**
// CHECK: store double* [[A:%.+]], double** [[CP0]], // CHECK: store double* [[A:%.+]], double** [[CP0]],
// CHECK: [[SZ0:%.+]] = getelementptr inbounds [6 x i{{[0-9]+}}], [6 x i{{[0-9]+}}]* [[SIZET4]], i{{[0-9]+}} 0, i{{[0-9]+}} 0 // CHECK: [[SZ0:%.+]] = getelementptr inbounds [5 x i{{[0-9]+}}], [5 x i{{[0-9]+}}]* [[SIZET4]], i{{[0-9]+}} 0, i{{[0-9]+}} 0
// CHECK: store i{{[0-9]+}} %{{.+}}, i{{[0-9]+}}* [[SZ0]], // CHECK: store i{{[0-9]+}} 8, i{{[0-9]+}}* [[SZ0]],
// CHECK: [[BP1:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BASE_PTRS4]], i{{[0-9]+}} 0, i{{[0-9]+}} 1
// CHECK: [[CBP1:%.+]] = bitcast i8** [[BP1]] to %struct.S1**
// CHECK: store %struct.S1* [[THIS]], %struct.S1** [[CBP1]],
// CHECK: [[P1:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[PTRS4]], i{{[0-9]+}} 0, i{{[0-9]+}} 1
// CHECK: [[CP1:%.+]] = bitcast i8** [[P1]] to double**
// CHECK: store double* [[A]], double** [[CP1]],
// CHECK: [[SZ1:%.+]] = getelementptr inbounds [6 x i{{[0-9]+}}], [6 x i{{[0-9]+}}]* [[SIZET4]], i{{[0-9]+}} 0, i{{[0-9]+}} 1
// CHECK: store i{{[0-9]+}} 8, i{{[0-9]+}}* [[SZ1]],
// firstprivate(b): base_ptr = b, ptr = b, size = 4 (pass by-value) // firstprivate(b): base_ptr = b, ptr = b, size = 4 (pass by-value)
// CHECK: [[BASE_PTRS_GEP4_1:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BASE_PTRS4]], i{{[0-9]+}} 0, i{{[0-9]+}} 2 // CHECK: [[BASE_PTRS_GEP4_1:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BASE_PTRS4]], i{{[0-9]+}} 0, i{{[0-9]+}} 1
// CHECK: [[BCAST_TOPTR:%.+]] = bitcast i8** [[BASE_PTRS_GEP4_1]] to i{{[0-9]+}}* // CHECK: [[BCAST_TOPTR:%.+]] = bitcast i8** [[BASE_PTRS_GEP4_1]] to i{{[0-9]+}}*
// CHECK: store i{{[0-9]+}} [[B_CAST:%.+]], i{{[0-9]+}}* [[BCAST_TOPTR]], // CHECK: store i{{[0-9]+}} [[B_CAST:%.+]], i{{[0-9]+}}* [[BCAST_TOPTR]],
// CHECK: [[PTRS_GEP4_1:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[PTRS4]], i{{[0-9]+}} 0, i{{[0-9]+}} 2 // CHECK: [[PTRS_GEP4_1:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[PTRS4]], i{{[0-9]+}} 0, i{{[0-9]+}} 1
// CHECK: [[BCAST_TOPTR:%.+]] = bitcast i8** [[PTRS_GEP4_1]] to i{{[0-9]+}}* // CHECK: [[BCAST_TOPTR:%.+]] = bitcast i8** [[PTRS_GEP4_1]] to i{{[0-9]+}}*
// CHECK: store i{{[0-9]+}} [[B_CAST]], i{{[0-9]+}}* [[BCAST_TOPTR]], // CHECK: store i{{[0-9]+}} [[B_CAST]], i{{[0-9]+}}* [[BCAST_TOPTR]],
// CHECK: [[SIZES_GEP4_1:%.+]] = getelementptr inbounds [6 x i{{[0-9]+}}], [6 x i{{[0-9]+}}]* [[SIZET4]], i{{[0-9]+}} 0, i{{[0-9]+}} 2 // CHECK: [[SIZES_GEP4_1:%.+]] = getelementptr inbounds [5 x i{{[0-9]+}}], [5 x i{{[0-9]+}}]* [[SIZET4]], i{{[0-9]+}} 0, i{{[0-9]+}} 1
// CHECK: store i{{[0-9]+}} 4, i{{[0-9]+}}* [[SIZES_GEP4_1]], // CHECK: store i{{[0-9]+}} 4, i{{[0-9]+}}* [[SIZES_GEP4_1]],
// firstprivate(c), 3 entries: 2, n, c // firstprivate(c), 3 entries: 2, n, c
// CHECK: [[BASE_PTRS_GEP4_2:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BASE_PTRS4]], i{{[0-9]+}} 0, i{{[0-9]+}} 3 // CHECK: [[BASE_PTRS_GEP4_2:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BASE_PTRS4]], i{{[0-9]+}} 0, i{{[0-9]+}} 2
// CHECK: [[BCAST_TOPTR:%.+]] = bitcast i8** [[BASE_PTRS_GEP4_2]] to i{{[0-9]+}}* // CHECK: [[BCAST_TOPTR:%.+]] = bitcast i8** [[BASE_PTRS_GEP4_2]] to i{{[0-9]+}}*
// CHECK: store i{{[0-9]+}} 2, i{{[0-9]+}}* [[BCAST_TOPTR]], // CHECK: store i{{[0-9]+}} 2, i{{[0-9]+}}* [[BCAST_TOPTR]],
// CHECK: [[PTRS_GEP4_2:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[PTRS4]], i{{[0-9]+}} 0, i{{[0-9]+}} 3 // CHECK: [[PTRS_GEP4_2:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[PTRS4]], i{{[0-9]+}} 0, i{{[0-9]+}} 2
// CHECK: [[BCAST_TOPTR:%.+]] = bitcast i8** [[PTRS_GEP4_2]] to i{{[0-9]+}}* // CHECK: [[BCAST_TOPTR:%.+]] = bitcast i8** [[PTRS_GEP4_2]] to i{{[0-9]+}}*
// CHECK: store i{{[0-9]+}} 2, i{{[0-9]+}}* [[BCAST_TOPTR]], // CHECK: store i{{[0-9]+}} 2, i{{[0-9]+}}* [[BCAST_TOPTR]],
// CHECK: [[SIZES_GEP4_2:%.+]] = getelementptr inbounds [6 x i{{[0-9]+}}], [6 x i{{[0-9]+}}]* [[SIZET4]], i{{[0-9]+}} 0, i{{[0-9]+}} 3 // CHECK: [[SIZES_GEP4_2:%.+]] = getelementptr inbounds [5 x i{{[0-9]+}}], [5 x i{{[0-9]+}}]* [[SIZET4]], i{{[0-9]+}} 0, i{{[0-9]+}} 2
// CHECK-64: store i{{[0-9]+}} 8, i{{[0-9]+}}* [[SIZES_GEP4_2]], // CHECK-64: store i{{[0-9]+}} 8, i{{[0-9]+}}* [[SIZES_GEP4_2]],
// CHECK-32: store i{{[0-9]+}} 4, i{{[0-9]+}}* [[SIZES_GEP4_2]], // CHECK-32: store i{{[0-9]+}} 4, i{{[0-9]+}}* [[SIZES_GEP4_2]],
// CHECK: [[BASE_PTRS_GEP4_3:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BASE_PTRS4]], i{{[0-9]+}} 0, i{{[0-9]+}} 4 // CHECK: [[BASE_PTRS_GEP4_3:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BASE_PTRS4]], i{{[0-9]+}} 0, i{{[0-9]+}} 3
// CHECK: [[BCAST_TOPTR:%.+]] = bitcast i8** [[BASE_PTRS_GEP4_3]] to i{{[0-9]+}}* // CHECK: [[BCAST_TOPTR:%.+]] = bitcast i8** [[BASE_PTRS_GEP4_3]] to i{{[0-9]+}}*
// CHECK: store i{{[0-9]+}} [[N:%.+]], i{{[0-9]+}}* [[BCAST_TOPTR]], // CHECK: store i{{[0-9]+}} [[N:%.+]], i{{[0-9]+}}* [[BCAST_TOPTR]],
// CHECK: [[PTRS_GEP4_3:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[PTRS4]], i{{[0-9]+}} 0, i{{[0-9]+}} 4 // CHECK: [[PTRS_GEP4_3:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[PTRS4]], i{{[0-9]+}} 0, i{{[0-9]+}} 3
// CHECK: [[BCAST_TOPTR:%.+]] = bitcast i8** [[PTRS_GEP4_3]] to i{{[0-9]+}}* // CHECK: [[BCAST_TOPTR:%.+]] = bitcast i8** [[PTRS_GEP4_3]] to i{{[0-9]+}}*
// CHECK: store i{{[0-9]+}} [[N]], i{{[0-9]+}}* [[BCAST_TOPTR]], // CHECK: store i{{[0-9]+}} [[N]], i{{[0-9]+}}* [[BCAST_TOPTR]],
// CHECK: [[SIZES_GEP4_3:%.+]] = getelementptr inbounds [6 x i{{[0-9]+}}], [6 x i{{[0-9]+}}]* [[SIZET4]], i{{[0-9]+}} 0, i{{[0-9]+}} 4 // CHECK: [[SIZES_GEP4_3:%.+]] = getelementptr inbounds [5 x i{{[0-9]+}}], [5 x i{{[0-9]+}}]* [[SIZET4]], i{{[0-9]+}} 0, i{{[0-9]+}} 3
// CHECK-64: store i{{[0-9]+}} 8, i{{[0-9]+}}* [[SIZES_GEP4_3]], // CHECK-64: store i{{[0-9]+}} 8, i{{[0-9]+}}* [[SIZES_GEP4_3]],
// CHECK-32: store i{{[0-9]+}} 4, i{{[0-9]+}}* [[SIZES_GEP4_3]], // CHECK-32: store i{{[0-9]+}} 4, i{{[0-9]+}}* [[SIZES_GEP4_3]],
// CHECK: [[BASE_PTRS_GEP4_4:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BASE_PTRS4]], i{{[0-9]+}} 0, i{{[0-9]+}} 5 // CHECK: [[BASE_PTRS_GEP4_4:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BASE_PTRS4]], i{{[0-9]+}} 0, i{{[0-9]+}} 4
// CHECK: [[BCAST_TOPTR:%.+]] = bitcast i8** [[BASE_PTRS_GEP4_4]] to i{{[0-9]+}}** // CHECK: [[BCAST_TOPTR:%.+]] = bitcast i8** [[BASE_PTRS_GEP4_4]] to i{{[0-9]+}}**
// CHECK: store i{{[0-9]+}}* [[B:%.+]], i{{[0-9]+}}** [[BCAST_TOPTR]], // CHECK: store i{{[0-9]+}}* [[B:%.+]], i{{[0-9]+}}** [[BCAST_TOPTR]],
// CHECK: [[PTRS_GEP4_4:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[PTRS4]], i{{[0-9]+}} 0, i{{[0-9]+}} 5 // CHECK: [[PTRS_GEP4_4:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[PTRS4]], i{{[0-9]+}} 0, i{{[0-9]+}} 4
// CHECK: [[BCAST_TOPTR:%.+]] = bitcast i8** [[PTRS_GEP4_4]] to i{{[0-9]+}}** // CHECK: [[BCAST_TOPTR:%.+]] = bitcast i8** [[PTRS_GEP4_4]] to i{{[0-9]+}}**
// CHECK: store i{{[0-9]+}}* [[B]], i{{[0-9]+}}** [[BCAST_TOPTR]], // CHECK: store i{{[0-9]+}}* [[B]], i{{[0-9]+}}** [[BCAST_TOPTR]],
// CHECK: [[SIZES_GEP4_4:%.+]] = getelementptr inbounds [6 x i{{[0-9]+}}], [6 x i{{[0-9]+}}]* [[SIZET4]], i{{[0-9]+}} 0, i{{[0-9]+}} 5 // CHECK: [[SIZES_GEP4_4:%.+]] = getelementptr inbounds [5 x i{{[0-9]+}}], [5 x i{{[0-9]+}}]* [[SIZET4]], i{{[0-9]+}} 0, i{{[0-9]+}} 4
// CHECK: store i{{[0-9]+}} [[B_SIZE:%.+]], i{{[0-9]+}}* [[SIZES_GEP4_4]], // CHECK: store i{{[0-9]+}} [[B_SIZE:%.+]], i{{[0-9]+}}* [[SIZES_GEP4_4]],
// only check that we use the map types stored in the global variable // only check that we use the map types stored in the global variable
// CHECK: call i32 @__tgt_target_mapper(i64 -1, {{.+}}, i32 6, i8** {{.+}}, i8** {{.+}}, i{{[0-9]+}}* {{.+}}, i64* getelementptr inbounds ([6 x i64], [6 x i64]* [[MAPT4]], i32 0, i32 0), i8** null) // CHECK: call i32 @__tgt_target_mapper(i64 -1, {{.+}}, i32 5, i8** {{.+}}, i8** {{.+}}, i{{[0-9]+}}* {{.+}}, i64* getelementptr inbounds ([5 x i64], [5 x i64]* [[MAPT4]], i32 0, i32 0), i8** null)
// TCHECK: define weak void @__omp_offloading_{{.+}}([[S1]]* [[TH:%.+]], i{{[0-9]+}} [[B_IN:%.+]], i{{[0-9]+}} [[VLA:%.+]], i{{[0-9]+}} [[VLA1:%.+]], i{{[0-9]+}}{{.+}} [[C_IN:%.+]]) // TCHECK: define weak void @__omp_offloading_{{.+}}([[S1]]* [[TH:%.+]], i{{[0-9]+}} [[B_IN:%.+]], i{{[0-9]+}} [[VLA:%.+]], i{{[0-9]+}} [[VLA1:%.+]], i{{[0-9]+}}{{.+}} [[C_IN:%.+]])
// TCHECK: [[TH_ADDR:%.+]] = alloca [[S1]]*, // TCHECK: [[TH_ADDR:%.+]] = alloca [[S1]]*,
// TCHECK: [[B_ADDR:%.+]] = alloca i{{[0-9]+}}, // TCHECK: [[B_ADDR:%.+]] = alloca i{{[0-9]+}},
// TCHECK: [[VLA_ADDR:%.+]] = alloca i{{[0-9]+}}, // TCHECK: [[VLA_ADDR:%.+]] = alloca i{{[0-9]+}},
// TCHECK: [[VLA_ADDR2:%.+]] = alloca i{{[0-9]+}}, // TCHECK: [[VLA_ADDR2:%.+]] = alloca i{{[0-9]+}},
// TCHECK: [[C_ADDR:%.+]] = alloca i{{[0-9]+}}*, // TCHECK: [[C_ADDR:%.+]] = alloca i{{[0-9]+}}*,
// TCHECK-NOT: alloca i{{[0-9]+}}, // TCHECK-NOT: alloca i{{[0-9]+}},
▲ Show 20 Lines • Show All 122 Lines • Show Last 20 Lines

clang/test/OpenMP/target_map_codegen_20.cpp

Show First 20 Lines • Show All 58 Lines • ▼ Show 20 Lines
// RUN: %clang_cc1 -DCK21 -fopenmp-simd -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s // RUN: %clang_cc1 -DCK21 -fopenmp-simd -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s
// RUN: %clang_cc1 -fopenmp-simd -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck -allow-deprecated-dag-overlap --check-prefix SIMD-ONLY20 %s // RUN: %clang_cc1 -fopenmp-simd -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck -allow-deprecated-dag-overlap --check-prefix SIMD-ONLY20 %s
// SIMD-ONLY20-NOT: {{__kmpc|__tgt}} // SIMD-ONLY20-NOT: {{__kmpc|__tgt}}
#ifdef CK21 #ifdef CK21
// CK21: [[ST:%.+]] = type { i32, i32, float* } // CK21: [[ST:%.+]] = type { i32, i32, float* }
// CK21-LABEL: @.__omp_offloading_{{.*}}foo{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0 // CK21-LABEL: @.__omp_offloading_{{.*}}foo{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
// CK21-USE: [[MTYPE00:@.+]] = private {{.*}}constant [2 x i64] [i64 32, i64 281474976710659] // CK21: [[SIZE00:@.+]] = private {{.*}}constant [1 x i64] [i64 4]
// CK21-NOUSE: [[MTYPE00:@.+]] = private {{.*}}constant [2 x i64] [i64 0, i64 281474976710659] // CK21-USE: [[MTYPE00:@.+]] = private {{.*}}constant [1 x i64] [i64 35]
// CK21-NOUSE: [[MTYPE00:@.+]] = private {{.*}}constant [1 x i64] [i64 3]
// CK21-LABEL: @.__omp_offloading_{{.*}}foo{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0 // CK21-LABEL: @.__omp_offloading_{{.*}}foo{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
// CK21: [[SIZE01:@.+]] = private {{.*}}constant [1 x i64] [i64 492] // CK21: [[SIZE01:@.+]] = private {{.*}}constant [1 x i64] [i64 492]
// CK21-USE: [[MTYPE01:@.+]] = private {{.*}}constant [1 x i64] [i64 35] // CK21-USE: [[MTYPE01:@.+]] = private {{.*}}constant [1 x i64] [i64 35]
// CK21-NOUSE: [[MTYPE01:@.+]] = private {{.*}}constant [1 x i64] [i64 3] // CK21-NOUSE: [[MTYPE01:@.+]] = private {{.*}}constant [1 x i64] [i64 3]
// CK21-LABEL: @.__omp_offloading_{{.*}}foo{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0 // CK21-LABEL: @.__omp_offloading_{{.*}}foo{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
// CK21-USE: [[MTYPE02:@.+]] = private {{.*}}constant [2 x i64] [i64 32, i64 281474976710674] // CK21-USE: [[MTYPE02:@.+]] = private {{.*}}constant [2 x i64] [i64 32, i64 281474976710674]
Show All 21 Linesstruct CC {
int A2; int A2;
float *B; float *B;
int foo(T arg) { int foo(T arg) {
float la[X]; float la[X];
T *lb; T *lb;
// Region 00 // Region 00
// CK21-DAG: call i32 @__tgt_target_mapper(i64 {{[^,]+}}, i8* {{[^,]+}}, i32 2, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], i64* [[GEPS:%.+]], {{.+}}getelementptr {{.+}}[2 x i{{.+}}]* [[MTYPE00]]{{.+}}, i8** null) // CK21-DAG: call i32 @__tgt_target_mapper(i64 {{[^,]+}}, i8* {{[^,]+}}, i32 1, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], {{.+}}getelementptr {{.+}}[1 x i{{.+}}]* [[SIZE00]]{{.+}}, {{.+}}getelementptr {{.+}}[1 x i{{.+}}]* [[MTYPE00]]{{.+}}, i8** null)
// CK21-DAG: [[GEPBP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]] // CK21-DAG: [[GEPBP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
// CK21-DAG: [[GEPP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]] // CK21-DAG: [[GEPP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
// CK21-DAG: [[GEPS]] = getelementptr inbounds {{.+}}[[S:%[^,]+]]
// CK21-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0 // CK21-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
// CK21-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0 // CK21-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
// CK21-DAG: [[S0:%.+]] = getelementptr inbounds {{.+}}[[S]], i{{.+}} 0, i{{.+}} 0
// CK21-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[ST]]** // CK21-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[ST]]**
// CK21-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to i32** // CK21-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to i32**
// CK21-DAG: store [[ST]]* [[VAR0:%.+]], [[ST]]** [[CBP0]] // CK21-DAG: store [[ST]]* [[VAR0:%.+]], [[ST]]** [[CBP0]]
// CK21-DAG: store i32* [[SEC0:%.+]], i32** [[CP0]] // CK21-DAG: store i32* [[SEC0:%.+]], i32** [[CP0]]
// CK21-DAG: store i64 {{%.+}}, i64* [[S0]]
// CK21-DAG: [[SEC0]] = getelementptr {{.*}}[[ST]]* [[VAR0:%.+]], i{{.+}} 0, i{{.+}} 0 // CK21-DAG: [[SEC0]] = getelementptr {{.*}}[[ST]]* [[VAR0:%.+]], i{{.+}} 0, i{{.+}} 0
// CK21-DAG: [[BP1:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 1
// CK21-DAG: [[P1:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 1
// CK21-DAG: [[S1:%.+]] = getelementptr inbounds {{.+}}[[S]], i{{.+}} 0, i{{.+}} 1
// CK21-DAG: [[CBP1:%.+]] = bitcast i8** [[BP1]] to [[ST]]**
// CK21-DAG: [[CP1:%.+]] = bitcast i8** [[P1]] to i32**
// CK21-DAG: store [[ST]]* [[VAR1:%.+]], [[ST]]** [[CBP1]]
// CK21-DAG: store i32* [[SEC1:%.+]], i32** [[CP1]]
// CK21-DAG: store i64 {{.+}}, i64* [[S1]]
// CK21-DAG: [[SEC1]] = getelementptr {{.*}}[[ST]]* [[VAR1:%.+]], i{{.+}} 0, i{{.+}} 0
// CK21-USE: call void [[CALL00:@.+]]([[ST]]* {{[^,]+}}) // CK21-USE: call void [[CALL00:@.+]]([[ST]]* {{[^,]+}})
// CK21-NOUSE: call void [[CALL00:@.+]]() // CK21-NOUSE: call void [[CALL00:@.+]]()
#pragma omp target map(A) #pragma omp target map(A)
{ {
#ifdef USE #ifdef USE
A += 1; A += 1;
#endif #endif
} }
▲ Show 20 Lines • Show All 166 Lines • Show Last 20 Lines

clang/test/OpenMP/target_map_codegen_23.cpp

Show First 20 Lines • Show All 51 Lines • ▼ Show 20 Lines
struct SC{ struct SC{
int a; int a;
struct SB s; struct SB s;
struct SB *p; struct SB *p;
int b[10]; int b[10];
}; };
// CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0 // CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
// CK24: [[MTYPE01:@.+]] = private {{.*}}constant [2 x i64] [i64 32, i64 281474976710659] // CK24: [[SIZE01:@.+]] = private {{.*}}constant [1 x i64] [i64 4]
// CK24: [[MTYPE01:@.+]] = private {{.*}}constant [1 x i64] [i64 35]
// CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0 // CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
// CK24: [[MTYPE13:@.+]] = private {{.*}}constant [2 x i64] [i64 32, i64 281474976710659] // CK24: [[SIZE13:@.+]] = private {{.*}}constant [1 x i64] [i64 4]
// CK24: [[MTYPE13:@.+]] = private {{.*}}constant [1 x i64] [i64 35]
// CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0 // CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
// CK24: [[MTYPE14:@.+]] = private {{.*}}constant [2 x i64] [i64 32, i64 281474976710659] // CK24: [[SIZE14:@.+]] = private {{.*}}constant [1 x i64] [i64 {{48|56}}]
// CK24: [[MTYPE14:@.+]] = private {{.*}}constant [1 x i64] [i64 35]
// CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0 // CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
// CK24: [[MTYPE15:@.+]] = private {{.*}}constant [2 x i64] [i64 32, i64 281474976710659] // CK24: [[SIZE15:@.+]] = private {{.*}}constant [1 x i64] [i64 4]
// CK24: [[MTYPE15:@.+]] = private {{.*}}constant [1 x i64] [i64 35]
// CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0 // CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
// CK24: [[MTYPE16:@.+]] = private {{.*}}constant [2 x i64] [i64 32, i64 281474976710659] // CK24: [[MTYPE16:@.+]] = private {{.*}}constant [2 x i64] [i64 32, i64 281474976710659]
// CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0 // CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
// CK24: [[MTYPE17:@.+]] = private {{.*}}constant [2 x i64] [i64 32, i64 281474976710675] // CK24: [[MTYPE17:@.+]] = private {{.*}}constant [2 x i64] [i64 32, i64 281474976710675]
// CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0 // CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
// CK24: [[MTYPE18:@.+]] = private {{.*}}constant [2 x i64] [i64 32, i64 281474976710659] // CK24: [[SIZE18:@.+]] = private {{.*}}constant [1 x i64] [i64 4]
// CK24: [[MTYPE18:@.+]] = private {{.*}}constant [1 x i64] [i64 35]
// CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0 // CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
// CK24: [[MTYPE19:@.+]] = private {{.*}}constant [3 x i64] [i64 32, i64 281474976710659, i64 281474976710675] // CK24: [[MTYPE19:@.+]] = private {{.*}}constant [3 x i64] [i64 32, i64 281474976710659, i64 281474976710675]
// CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0 // CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
// CK24: [[MTYPE20:@.+]] = private {{.*}}constant [2 x i64] [i64 32, i64 281474976710675] // CK24: [[MTYPE20:@.+]] = private {{.*}}constant [2 x i64] [i64 32, i64 281474976710675]
// CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0 // CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
Show All 9 Lines
// CK24: [[MTYPE24:@.+]] = private {{.*}}constant [4 x i64] [i64 32, i64 281474976710672, i64 16, i64 19] // CK24: [[MTYPE24:@.+]] = private {{.*}}constant [4 x i64] [i64 32, i64 281474976710672, i64 16, i64 19]
// CK24-LABEL: explicit_maps_struct_fields // CK24-LABEL: explicit_maps_struct_fields
int explicit_maps_struct_fields(int a){ int explicit_maps_struct_fields(int a){
SC s; SC s;
SC *p; SC *p;
// Region 01 // Region 01
// CK24-DAG: call i32 @__tgt_target_mapper(i64 {{[^,]+}}, i8* {{[^,]+}}, i32 2, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], i64* [[GEPS:%.+]], {{.+}}getelementptr {{.+}}[2 x i{{.+}}]* [[MTYPE01]]{{.+}}, i8** null) // CK24-DAG: call i32 @__tgt_target_mapper(i64 {{[^,]+}}, i8* {{[^,]+}}, i32 1, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], {{.+}}getelementptr {{.+}}[1 x i{{.+}}]* [[SIZE01]]{{.+}}, {{.+}}getelementptr {{.+}}[1 x i{{.+}}]* [[MTYPE01]]{{.+}}, i8** null)
// CK24-DAG: [[GEPBP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]] // CK24-DAG: [[GEPBP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
// CK24-DAG: [[GEPP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]] // CK24-DAG: [[GEPP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
// CK24-DAG: [[GEPS]] = getelementptr inbounds {{.+}}[[S:%[^,]+]]
// CK24-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0 // CK24-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
// CK24-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0 // CK24-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
// CK24-DAG: [[S0:%.+]] = getelementptr inbounds {{.+}}[[S]], i{{.+}} 0, i{{.+}} 0
// CK24-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[SC]]** // CK24-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[SC]]**
// CK24-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to i32** // CK24-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to i32**
// CK24-DAG: store [[SC]]* [[VAR0:%.+]], [[SC]]** [[CBP0]] // CK24-DAG: store [[SC]]* [[VAR0:%.+]], [[SC]]** [[CBP0]]
// CK24-DAG: store i32* [[SEC0:%.+]], i32** [[CP0]] // CK24-DAG: store i32* [[SEC0:%.+]], i32** [[CP0]]
// CK24-DAG: store i64 {{%.+}}, i64* [[S0]]
// CK24-DAG: [[SEC0]] = getelementptr {{.*}}[[SC]]* [[VAR0]], i{{.+}} 0, i{{.+}} 0 // CK24-DAG: [[SEC0]] = getelementptr {{.*}}[[SC]]* [[VAR0]], i{{.+}} 0, i{{.+}} 0
// CK24-DAG: [[BP1:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 1
// CK24-DAG: [[P1:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 1
// CK24-DAG: [[S1:%.+]] = getelementptr inbounds {{.+}}[[S]], i{{.+}} 0, i{{.+}} 1
// CK24-DAG: [[CBP1:%.+]] = bitcast i8** [[BP1]] to [[SC]]**
// CK24-DAG: [[CP1:%.+]] = bitcast i8** [[P1]] to i32**
// CK24-DAG: store [[SC]]* [[VAR0]], [[SC]]** [[CBP1]]
// CK24-DAG: store i32* [[SEC0]], i32** [[CP1]]
// CK24-DAG: store i64 {{.+}}, i64* [[S1]]
// CK24: call void [[CALL01:@.+]]([[SC]]* {{[^,]+}}) // CK24: call void [[CALL01:@.+]]([[SC]]* {{[^,]+}})
#pragma omp target map(s.a) #pragma omp target map(s.a)
{ s.a++; } { s.a++; }
// //
// Same thing but starting from a pointer. // Same thing but starting from a pointer.
// //
// Region 13 // Region 13
// CK24-DAG: call i32 @__tgt_target_mapper(i64 {{[^,]+}}, i8* {{[^,]+}}, i32 2, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], i64* [[GEPS:%.+]], {{.+}}getelementptr {{.+}}[2 x i{{.+}}]* [[MTYPE13]]{{.+}}, i8** null) // CK24-DAG: call i32 @__tgt_target_mapper(i64 {{[^,]+}}, i8* {{[^,]+}}, i32 1, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], {{.+}}getelementptr {{.+}}[1 x i{{.+}}]* [[SIZE13]]{{.+}}, {{.+}}getelementptr {{.+}}[1 x i{{.+}}]* [[MTYPE13]]{{.+}}, i8** null)
// CK24-DAG: [[GEPBP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]] // CK24-DAG: [[GEPBP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
// CK24-DAG: [[GEPP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]] // CK24-DAG: [[GEPP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
// CK24-DAG: [[GEPS]] = getelementptr inbounds {{.+}}[[S:%[^,]+]]
// CK24-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0 // CK24-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
// CK24-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0 // CK24-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
// CK24-DAG: [[S0:%.+]] = getelementptr inbounds {{.+}}[[S]], i{{.+}} 0, i{{.+}} 0
// CK24-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[SC]]** // CK24-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[SC]]**
// CK24-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to i32** // CK24-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to i32**
// CK24-DAG: store [[SC]]* [[VAR0:%.+]], [[SC]]** [[CBP0]] // CK24-DAG: store [[SC]]* [[VAR0:%.+]], [[SC]]** [[CBP0]]
// CK24-DAG: store i32* [[SEC0:%.+]], i32** [[CP0]] // CK24-DAG: store i32* [[SEC0:%.+]], i32** [[CP0]]
// CK24-DAG: store i64 {{%.+}}, i64* [[S0]]
// CK24-DAG: [[SEC0]] = getelementptr {{.*}}[[SC]]* [[VAR00:%.+]], i{{.+}} 0, i{{.+}} 0 // CK24-DAG: [[SEC0]] = getelementptr {{.*}}[[SC]]* [[VAR00:%.+]], i{{.+}} 0, i{{.+}} 0
// CK24-DAG: [[BP1:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 1
// CK24-DAG: [[P1:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 1
// CK24-DAG: [[S1:%.+]] = getelementptr inbounds {{.+}}[[S]], i{{.+}} 0, i{{.+}} 1
// CK24-DAG: [[CBP1:%.+]] = bitcast i8** [[BP1]] to [[SC]]**
// CK24-DAG: [[CP1:%.+]] = bitcast i8** [[P1]] to i32**
// CK24-DAG: store [[SC]]* [[VAR0]], [[SC]]** [[CBP1]]
// CK24-DAG: store i32* [[SEC0]], i32** [[CP1]]
// CK24-DAG: store i64 {{.+}}, i64* [[S1]]
// CK24-DAG: [[VAR0]] = load [[SC]]*, [[SC]]** %{{.+}} // CK24-DAG: [[VAR0]] = load [[SC]]*, [[SC]]** %{{.+}}
// CK24-DAG: [[VAR00]] = load [[SC]]*, [[SC]]** %{{.+}} // CK24-DAG: [[VAR00]] = load [[SC]]*, [[SC]]** %{{.+}}
// CK24: call void [[CALL13:@.+]]([[SC]]* {{[^,]+}}) // CK24: call void [[CALL13:@.+]]([[SC]]* {{[^,]+}})
#pragma omp target map(p->a) #pragma omp target map(p->a)
{ p->a++; } { p->a++; }
// Region 14 // Region 14
// CK24-DAG: call i32 @__tgt_target_mapper(i64 {{[^,]+}}, i8* {{[^,]+}}, i32 2, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], i64* [[GEPS:%.+]], {{.+}}getelementptr {{.+}}[2 x i{{.+}}]* [[MTYPE14]]{{.+}}, i8** null) // CK24-DAG: call i32 @__tgt_target_mapper(i64 {{[^,]+}}, i8* {{[^,]+}}, i32 1, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], {{.+}}getelementptr {{.+}}[1 x i{{.+}}]* [[SIZE14]]{{.+}}, {{.+}}getelementptr {{.+}}[1 x i{{.+}}]* [[MTYPE14]]{{.+}}, i8** null)
// CK24-DAG: [[GEPBP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]] // CK24-DAG: [[GEPBP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
// CK24-DAG: [[GEPP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]] // CK24-DAG: [[GEPP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
// CK24-DAG: [[GEPS]] = getelementptr inbounds {{.+}}[[S:%[^,]+]]
// CK24-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0 // CK24-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
// CK24-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0 // CK24-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
// CK24-DAG: [[S0:%.+]] = getelementptr inbounds {{.+}}[[S]], i{{.+}} 0, i{{.+}} 0
// CK24-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[SC]]** // CK24-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[SC]]**
// CK24-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to [[SA]]** // CK24-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to [[SA]]**
// CK24-DAG: store [[SC]]* [[VAR0:%.+]], [[SC]]** [[CBP0]] // CK24-DAG: store [[SC]]* [[VAR0:%.+]], [[SC]]** [[CBP0]]
// CK24-DAG: store [[SA]]* [[SEC0:%.+]], [[SA]]** [[CP0]] // CK24-DAG: store [[SA]]* [[SEC0:%.+]], [[SA]]** [[CP0]]
// CK24-DAG: store i64 {{%.+}}, i64* [[S0]]
// CK24-DAG: [[SEC0]] = getelementptr {{.*}}[[SB]]* [[SEC00:%[^,]+]], i{{.+}} 0, i{{.+}} 1 // CK24-DAG: [[SEC0]] = getelementptr {{.*}}[[SB]]* [[SEC00:%[^,]+]], i{{.+}} 0, i{{.+}} 1
// CK24-DAG: [[SEC00]] = getelementptr {{.*}}[[SC]]* [[VAR00:%.+]], i{{.+}} 0, i{{.+}} 1 // CK24-DAG: [[SEC00]] = getelementptr {{.*}}[[SC]]* [[VAR00:%.+]], i{{.+}} 0, i{{.+}} 1
// CK24-DAG: [[BP1:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 1
// CK24-DAG: [[P1:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 1
// CK24-DAG: [[S1:%.+]] = getelementptr inbounds {{.+}}[[S]], i{{.+}} 0, i{{.+}} 1
// CK24-DAG: [[CBP1:%.+]] = bitcast i8** [[BP1]] to [[SC]]**
// CK24-DAG: [[CP1:%.+]] = bitcast i8** [[P1]] to [[SA]]**
// CK24-DAG: store [[SC]]* [[VAR0]], [[SC]]** [[CBP1]]
// CK24-DAG: store [[SA]]* [[SEC0]], [[SA]]** [[CP1]]
// CK24-DAG: store i64 {{.+}}, i64* [[S1]]
// CK24-DAG: [[VAR0]] = load [[SC]]*, [[SC]]** %{{.+}} // CK24-DAG: [[VAR0]] = load [[SC]]*, [[SC]]** %{{.+}}
// CK24-DAG: [[VAR00]] = load [[SC]]*, [[SC]]** %{{.+}} // CK24-DAG: [[VAR00]] = load [[SC]]*, [[SC]]** %{{.+}}
// CK24: call void [[CALL14:@.+]]([[SC]]* {{[^,]+}}) // CK24: call void [[CALL14:@.+]]([[SC]]* {{[^,]+}})
#pragma omp target map(p->s.s) #pragma omp target map(p->s.s)
{ p->a++; } { p->a++; }
// Region 15 // Region 15
// CK24-DAG: call i32 @__tgt_target_mapper(i64 {{[^,]+}}, i8* {{[^,]+}}, i32 2, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], i[[sz:64|32]]* [[GEPS:%.+]], {{.+}}getelementptr {{.+}}[2 x i{{.+}}]* [[MTYPE15]]{{.+}}, i8** null) // CK24-DAG: call i32 @__tgt_target_mapper(i64 {{[^,]+}}, i8* {{[^,]+}}, i32 1, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], {{.+}}getelementptr {{.+}}[1 x i{{.+}}]* [[SIZE15]]{{.+}}, {{.+}}getelementptr {{.+}}[1 x i{{.+}}]* [[MTYPE15]]{{.+}}, i8** null)
// CK24-DAG: [[GEPBP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]] // CK24-DAG: [[GEPBP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
// CK24-DAG: [[GEPP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]] // CK24-DAG: [[GEPP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
// CK24-DAG: [[GEPS]] = getelementptr inbounds {{.+}}[[S:%[^,]+]]
// CK24-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0 // CK24-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
// CK24-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0 // CK24-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
// CK24-DAG: [[S0:%.+]] = getelementptr inbounds {{.+}}[[S]], i{{.+}} 0, i{{.+}} 0
// CK24-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[SC]]** // CK24-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[SC]]**
// CK24-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to i32** // CK24-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to i32**
// CK24-DAG: store [[SC]]* [[VAR0:%.+]], [[SC]]** [[CBP0]] // CK24-DAG: store [[SC]]* [[VAR0:%.+]], [[SC]]** [[CBP0]]
// CK24-DAG: store i32* [[SEC0:%.+]], i32** [[CP0]] // CK24-DAG: store i32* [[SEC0:%.+]], i32** [[CP0]]
// CK24-DAG: store i64 {{%.+}}, i64* [[S0]]
// CK24-DAG: [[SEC0]] = getelementptr {{.*}}[[SA]]* [[SEC00:%[^,]+]], i{{.+}} 0, i{{.+}} 0 // CK24-DAG: [[SEC0]] = getelementptr {{.*}}[[SA]]* [[SEC00:%[^,]+]], i{{.+}} 0, i{{.+}} 0
// CK24-DAG: [[SEC00]] = getelementptr {{.*}}[[SB]]* [[SEC000:%[^,]+]], i{{.+}} 0, i{{.+}} 1 // CK24-DAG: [[SEC00]] = getelementptr {{.*}}[[SB]]* [[SEC000:%[^,]+]], i{{.+}} 0, i{{.+}} 1
// CK24-DAG: [[SEC000]] = getelementptr {{.*}}[[SC]]* [[VAR00:%.+]], i{{.+}} 0, i{{.+}} 1 // CK24-DAG: [[SEC000]] = getelementptr {{.*}}[[SC]]* [[VAR00:%.+]], i{{.+}} 0, i{{.+}} 1
// CK24-DAG: [[BP1:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 1
// CK24-DAG: [[P1:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 1
// CK24-DAG: [[S1:%.+]] = getelementptr inbounds {{.+}}[[S]], i{{.+}} 0, i{{.+}} 1
// CK24-DAG: [[CBP1:%.+]] = bitcast i8** [[BP1]] to [[SC]]**
// CK24-DAG: [[CP1:%.+]] = bitcast i8** [[P1]] to i32**
// CK24-DAG: store [[SC]]* [[VAR0]], [[SC]]** [[CBP1]]
// CK24-DAG: store i32* [[SEC0]], i32** [[CP1]]
// CK24-DAG: store i64 {{.+}}, i64* [[S1]]
// CK24-DAG: [[VAR0]] = load [[SC]]*, [[SC]]** %{{.+}} // CK24-DAG: [[VAR0]] = load [[SC]]*, [[SC]]** %{{.+}}
// CK24-DAG: [[VAR00]] = load [[SC]]*, [[SC]]** %{{.+}} // CK24-DAG: [[VAR00]] = load [[SC]]*, [[SC]]** %{{.+}}
// CK24: call void [[CALL15:@.+]]([[SC]]* {{[^,]+}}) // CK24: call void [[CALL15:@.+]]([[SC]]* {{[^,]+}})
#pragma omp target map(p->s.s.a) #pragma omp target map(p->s.s.a)
{ p->a++; } { p->a++; }
// Region 16 // Region 16
▲ Show 20 Lines • Show All 61 Lines • ▼ Show 20 Lines
// CK24-DAG: [[VAR00]] = load [[SC]]*, [[SC]]** %{{.+}} // CK24-DAG: [[VAR00]] = load [[SC]]*, [[SC]]** %{{.+}}
// CK24-DAG: [[VAR000]] = load [[SC]]*, [[SC]]** %{{.+}} // CK24-DAG: [[VAR000]] = load [[SC]]*, [[SC]]** %{{.+}}
// CK24: call void [[CALL17:@.+]]([[SC]]* {{[^,]+}}) // CK24: call void [[CALL17:@.+]]([[SC]]* {{[^,]+}})
#pragma omp target map(p->p[:5]) #pragma omp target map(p->p[:5])
{ p->a++; } { p->a++; }
// Region 18 // Region 18
// CK24-DAG: call i32 @__tgt_target_mapper(i64 {{[^,]+}}, i8* {{[^,]+}}, i32 2, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], i64* [[GEPS:%.+]], {{.+}}getelementptr {{.+}}[2 x i{{.+}}]* [[MTYPE18]]{{.+}}, i8** null) // CK24-DAG: call i32 @__tgt_target_mapper(i64 {{[^,]+}}, i8* {{[^,]+}}, i32 1, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], {{.+}}getelementptr {{.+}}[1 x i{{.+}}]* [[SIZE18]]{{.+}}, {{.+}}getelementptr {{.+}}[1 x i{{.+}}]* [[MTYPE18]]{{.+}}, i8** null)
// CK24-DAG: [[GEPBP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]] // CK24-DAG: [[GEPBP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
// CK24-DAG: [[GEPP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]] // CK24-DAG: [[GEPP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
// CK24-DAG: [[GEPS]] = getelementptr inbounds {{.+}}[[S:%[^,]+]]
// CK24-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0 // CK24-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
// CK24-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0 // CK24-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
// CK24-DAG: [[S0:%.+]] = getelementptr inbounds {{.+}}[[S]], i{{.+}} 0, i{{.+}} 0
// CK24-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[SC]]** // CK24-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[SC]]**
// CK24-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to i32** // CK24-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to i32**
// CK24-DAG: store [[SC]]* [[VAR0:%.+]], [[SC]]** [[CBP0]] // CK24-DAG: store [[SC]]* [[VAR0:%.+]], [[SC]]** [[CBP0]]
// CK24-DAG: store i32* [[SEC0:%.+]], i32** [[CP0]] // CK24-DAG: store i32* [[SEC0:%.+]], i32** [[CP0]]
// CK24-DAG: store i64 {{%.+}}, i64* [[S0]]
// CK24-DAG: [[SEC0]] = getelementptr {{.*}}[[SA]]* [[SEC00:%[^,]+]], i{{.+}} 0, i{{.+}} 0 // CK24-DAG: [[SEC0]] = getelementptr {{.*}}[[SA]]* [[SEC00:%[^,]+]], i{{.+}} 0, i{{.+}} 0
// CK24-DAG: [[SEC00]] = getelementptr {{.*}}[10 x [[SA]]]* [[SEC000:%[^,]+]], i{{.+}} 0, i{{.+}} 3 // CK24-DAG: [[SEC00]] = getelementptr {{.*}}[10 x [[SA]]]* [[SEC000:%[^,]+]], i{{.+}} 0, i{{.+}} 3
// CK24-DAG: [[SEC000]] = getelementptr {{.*}}[[SB]]* [[SEC0000:%[^,]+]], i{{.+}} 0, i{{.+}} 2 // CK24-DAG: [[SEC000]] = getelementptr {{.*}}[[SB]]* [[SEC0000:%[^,]+]], i{{.+}} 0, i{{.+}} 2
// CK24-DAG: [[SEC0000]] = getelementptr {{.*}}[[SC]]* [[VAR00:%.+]], i{{.+}} 0, i{{.+}} 1 // CK24-DAG: [[SEC0000]] = getelementptr {{.*}}[[SC]]* [[VAR00:%.+]], i{{.+}} 0, i{{.+}} 1
// CK24-DAG: [[BP1:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 1
// CK24-DAG: [[P1:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 1
// CK24-DAG: [[S1:%.+]] = getelementptr inbounds {{.+}}[[S]], i{{.+}} 0, i{{.+}} 1
// CK24-DAG: [[CBP1:%.+]] = bitcast i8** [[BP1]] to [[SC]]**
// CK24-DAG: [[CP1:%.+]] = bitcast i8** [[P1]] to i32**
// CK24-DAG: store [[SC]]* [[VAR0]], [[SC]]** [[CBP1]]
// CK24-DAG: store i32* [[SEC0]], i32** [[CP1]]
// CK24-DAG: store i64 {{.+}}, i64* [[S1]]
// CK24-DAG: [[VAR0]] = load [[SC]]*, [[SC]]** %{{.+}} // CK24-DAG: [[VAR0]] = load [[SC]]*, [[SC]]** %{{.+}}
// CK24-DAG: [[VAR00]] = load [[SC]]*, [[SC]]** %{{.+}} // CK24-DAG: [[VAR00]] = load [[SC]]*, [[SC]]** %{{.+}}
// CK24: call void [[CALL18:@.+]]([[SC]]* {{[^,]+}}) // CK24: call void [[CALL18:@.+]]([[SC]]* {{[^,]+}})
#pragma omp target map(p->s.sa[3].a) #pragma omp target map(p->s.sa[3].a)
{ p->a++; } { p->a++; }
// Region 19 // Region 19
▲ Show 20 Lines • Show All 296 Lines • Show Last 20 Lines

clang/test/OpenMP/target_map_codegen_24.cpp

Show All 31 Lines
// RUN: %clang_cc1 -fopenmp-simd -std=c++11 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck -allow-deprecated-dag-overlap --check-prefix SIMD-ONLY24 %s // RUN: %clang_cc1 -fopenmp-simd -std=c++11 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck -allow-deprecated-dag-overlap --check-prefix SIMD-ONLY24 %s
// SIMD-ONLY24-NOT: {{__kmpc|__tgt}} // SIMD-ONLY24-NOT: {{__kmpc|__tgt}}
#ifdef CK25 #ifdef CK25
// CK25: [[ST:%.+]] = type { i32, float } // CK25: [[ST:%.+]] = type { i32, float }
// CK25: [[CA00:%.+]] = type { [[ST]]* } // CK25: [[CA00:%.+]] = type { [[ST]]* }
// CK25: [[CA01:%.+]] = type { i32* } // CK25: [[CA01:%.+]] = type { i32* }
// CK25-LABEL: @.__omp_offloading_{{.*}}foo{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0 // CK25-LABEL: @.__omp_offloading_{{.*}}foo{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
// CK25: [[MTYPE00:@.+]] = private {{.*}}constant [2 x i64] [i64 32, i64 281474976710657] // CK25: [[SIZE00:@.+]] = private {{.*}}constant [1 x i64] [i64 4]
// CK25: [[MTYPE00:@.+]] = private {{.*}}constant [1 x i64] [i64 33]
// CK25-LABEL: @.__omp_offloading_{{.*}}foo{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0 // CK25-LABEL: @.__omp_offloading_{{.*}}foo{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
// CK25: [[SIZE01:@.+]] = private {{.*}}constant [1 x i64] [i64 4] // CK25: [[SIZE01:@.+]] = private {{.*}}constant [1 x i64] [i64 4]
// CK25: [[MTYPE01:@.+]] = private {{.*}}constant [1 x i64] [i64 33] // CK25: [[MTYPE01:@.+]] = private {{.*}}constant [1 x i64] [i64 33]
// CK25-LABEL: explicit_maps_with_inner_lambda{{.*}}( // CK25-LABEL: explicit_maps_with_inner_lambda{{.*}}(
template <int X, typename T> template <int X, typename T>
struct CC { struct CC {
T A; T A;
float B; float B;
int foo(T arg) { int foo(T arg) {
// Region 00 // Region 00
// CK25-DAG: call i32 @__tgt_target_mapper(i64 {{[^,]+}}, i8* {{[^,]+}}, i32 2, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], i64* [[GEPS:%.+]], {{.+}}getelementptr {{.+}}[2 x i{{.+}}]* [[MTYPE00]]{{.+}}, i8** null) // CK25-DAG: call i32 @__tgt_target_mapper(i64 {{[^,]+}}, i8* {{[^,]+}}, i32 1, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], {{.+}}getelementptr {{.+}}[1 x i{{.+}}]* [[SIZE00]]{{.+}}, {{.+}}getelementptr {{.+}}[1 x i{{.+}}]* [[MTYPE00]]{{.+}}, i8** null)
// CK25-DAG: [[GEPBP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]] // CK25-DAG: [[GEPBP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
// CK25-DAG: [[GEPP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]] // CK25-DAG: [[GEPP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
// CK25-DAG: [[GEPS]] = getelementptr inbounds {{.+}}[[S:%[^,]+]]
// CK25-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0 // CK25-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
// CK25-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0 // CK25-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
// CK25-DAG: [[S0:%.+]] = getelementptr inbounds {{.+}}[[S]], i{{.+}} 0, i{{.+}} 0
// CK25-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[ST]]** // CK25-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[ST]]**
// CK25-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to i32** // CK25-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to i32**
// CK25-DAG: store [[ST]]* [[VAR0:%.+]], [[ST]]** [[CBP0]] // CK25-DAG: store [[ST]]* [[VAR0:%.+]], [[ST]]** [[CBP0]]
// CK25-DAG: store i32* [[SEC0:%.+]], i32** [[CP0]] // CK25-DAG: store i32* [[SEC0:%.+]], i32** [[CP0]]
// CK25-DAG: store i64 {{%.+}}, i64* [[S0]]
// CK25-DAG: [[SEC0]] = getelementptr {{.*}}[[ST]]* [[VAR0:%.+]], i{{.+}} 0, i{{.+}} 0 // CK25-DAG: [[SEC0]] = getelementptr {{.*}}[[ST]]* [[VAR0:%.+]], i{{.+}} 0, i{{.+}} 0
// CK25-DAG: [[BP1:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 1
// CK25-DAG: [[P1:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 1
// CK25-DAG: [[S1:%.+]] = getelementptr inbounds {{.+}}[[S]], i{{.+}} 0, i{{.+}} 1
// CK25-DAG: [[CBP1:%.+]] = bitcast i8** [[BP1]] to [[ST]]**
// CK25-DAG: [[CP1:%.+]] = bitcast i8** [[P1]] to i32**
// CK25-DAG: store [[ST]]* [[VAR0]], [[ST]]** [[CBP1]]
// CK25-DAG: store i32* [[SEC0]], i32** [[CP1]]
// CK25-DAG: store i64 {{.+}}, i64* [[S1]]
// CK25: call void [[CALL00:@.+]]([[ST]]* {{[^,]+}}) // CK25: call void [[CALL00:@.+]]([[ST]]* {{[^,]+}})
#pragma omp target map(to:A) #pragma omp target map(to:A)
{ {
[&]() { [&]() {
A += 1; A += 1;
}(); }();
} }
▲ Show 20 Lines • Show All 48 Lines • Show Last 20 Lines

clang/test/OpenMP/target_parallel_codegen.cpp

Show First 20 Lines • Show All 93 Lines • ▼ Show 20 Lines
// CHECK-DAG: [[MAPT2:@.+]] = private unnamed_addr constant [1 x i64] [i64 800] // CHECK-DAG: [[MAPT2:@.+]] = private unnamed_addr constant [1 x i64] [i64 800]
// CHECK-DAG: [[SIZET3:@.+]] = private unnamed_addr constant [2 x i64] [i64 4, i64 2] // CHECK-DAG: [[SIZET3:@.+]] = private unnamed_addr constant [2 x i64] [i64 4, i64 2]
// CHECK-DAG: [[MAPT3:@.+]] = private unnamed_addr constant [2 x i64] [i64 800, i64 800] // CHECK-DAG: [[MAPT3:@.+]] = private unnamed_addr constant [2 x i64] [i64 800, i64 800]
// CHECK-DAG: [[MAPT4:@.+]] = private unnamed_addr constant [9 x i64] [i64 800, i64 547, i64 800, i64 547, i64 547, i64 800, i64 800, i64 547, i64 547] // CHECK-DAG: [[MAPT4:@.+]] = private unnamed_addr constant [9 x i64] [i64 800, i64 547, i64 800, i64 547, i64 547, i64 800, i64 800, i64 547, i64 547]
// CHECK-DAG: [[SIZET5:@.+]] = private unnamed_addr constant [3 x i64] [i64 4, i64 2, i64 40] // CHECK-DAG: [[SIZET5:@.+]] = private unnamed_addr constant [3 x i64] [i64 4, i64 2, i64 40]
// CHECK-DAG: [[MAPT5:@.+]] = private unnamed_addr constant [3 x i64] [i64 800, i64 800, i64 547] // CHECK-DAG: [[MAPT5:@.+]] = private unnamed_addr constant [3 x i64] [i64 800, i64 800, i64 547]
// CHECK-DAG: [[SIZET6:@.+]] = private unnamed_addr constant [4 x i64] [i64 4, i64 2, i64 1, i64 40] // CHECK-DAG: [[SIZET6:@.+]] = private unnamed_addr constant [4 x i64] [i64 4, i64 2, i64 1, i64 40]
// CHECK-DAG: [[MAPT6:@.+]] = private unnamed_addr constant [4 x i64] [i64 800, i64 800, i64 800, i64 547] // CHECK-DAG: [[MAPT6:@.+]] = private unnamed_addr constant [4 x i64] [i64 800, i64 800, i64 800, i64 547]
// CHECK-DAG: [[MAPT7:@.+]] = private unnamed_addr constant [6 x i64] [i64 32, i64 281474976711171, i64 800, i64 800, i64 800, i64 547] // CHECK-DAG: [[MAPT7:@.+]] = private unnamed_addr constant [5 x i64] [i64 547, i64 800, i64 800, i64 800, i64 547]
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
▲ Show 20 Lines • Show All 467 Lines • ▼ Show 20 Lines
// CHECK: br i1 [[IF]], label %[[TRY:[^,]+]], label %[[FAIL:[^,]+]] // CHECK: br i1 [[IF]], label %[[TRY:[^,]+]], label %[[FAIL:[^,]+]]
// CHECK: [[TRY]] // CHECK: [[TRY]]
// We capture 2 VLA sizes in this target region // We capture 2 VLA sizes in this target region
// CHECK: [[CELEMSIZE2:%.+]] = mul nuw i[[SZ]] 2, [[VLA0:%.+]] // CHECK: [[CELEMSIZE2:%.+]] = mul nuw i[[SZ]] 2, [[VLA0:%.+]]
// CHECK-64: [[CSIZE:%.+]] = mul nuw i64 [[CELEMSIZE2]], 2 // CHECK-64: [[CSIZE:%.+]] = mul nuw i64 [[CELEMSIZE2]], 2
// CHECK-32: [[CSZSIZE:%.+]] = mul nuw i32 [[CELEMSIZE2]], 2 // CHECK-32: [[CSZSIZE:%.+]] = mul nuw i32 [[CELEMSIZE2]], 2
// CHECK-32: [[CSIZE:%.+]] = sext i32 [[CSZSIZE]] to i64 // CHECK-32: [[CSIZE:%.+]] = sext i32 [[CSZSIZE]] to i64
// CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 6, i8** [[BPR:%[^,]+]], i8** [[PR:%[^,]+]], i64* [[SR:%[^,]+]], i64* getelementptr inbounds ([6 x i64], [6 x i64]* [[MAPT7]], i32 0, i32 0), i8** null, i32 1, i32 0) // CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 5, i8** [[BPR:%[^,]+]], i8** [[PR:%[^,]+]], i64* [[SR:%[^,]+]], i64* getelementptr inbounds ([5 x i64], [5 x i64]* [[MAPT7]], i32 0, i32 0), i8** null, i32 1, i32 0)
// CHECK-DAG: [[BPR]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP:%.+]], i32 0, i32 0 // CHECK-DAG: [[BPR]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP:%.+]], i32 0, i32 0
// CHECK-DAG: [[PR]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P:%.+]], i32 0, i32 0 // CHECK-DAG: [[PR]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P:%.+]], i32 0, i32 0
// CHECK-DAG: [[SR]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S:%.+]], i32 0, i32 0 // CHECK-DAG: [[SR]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S:%.+]], i32 0, i32 0
// CHECK-DAG: [[SADDR0:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX0:[0-9]+]] // CHECK-DAG: [[SADDR0:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX0:[0-9]+]]
// CHECK-DAG: [[BPADDR0:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX0]] // CHECK-DAG: [[BPADDR0:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX0]]
// CHECK-DAG: [[PADDR0:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX0]] // CHECK-DAG: [[PADDR0:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX0]]
// CHECK-DAG: [[SADDR1:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX1:[0-9]+]] // CHECK-DAG: [[SADDR1:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX1:[0-9]+]]
// CHECK-DAG: [[BPADDR1:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX1]] // CHECK-DAG: [[BPADDR1:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX1]]
// CHECK-DAG: [[PADDR1:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX1]] // CHECK-DAG: [[PADDR1:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX1]]
// CHECK-DAG: [[SADDR2:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX2:[0-9]+]] // CHECK-DAG: [[SADDR2:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX2:[0-9]+]]
// CHECK-DAG: [[BPADDR2:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX2]] // CHECK-DAG: [[BPADDR2:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX2]]
// CHECK-DAG: [[PADDR2:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX2]] // CHECK-DAG: [[PADDR2:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX2]]
// CHECK-DAG: [[SADDR3:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX3:[0-9]+]] // CHECK-DAG: [[SADDR3:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX3:[0-9]+]]
// CHECK-DAG: [[BPADDR3:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX3]] // CHECK-DAG: [[BPADDR3:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX3]]
// CHECK-DAG: [[PADDR3:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX3]] // CHECK-DAG: [[PADDR3:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX3]]
// CHECK-DAG: [[SADDR4:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX4:[0-9]+]] // CHECK-DAG: [[SADDR4:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX4:[0-9]+]]
// CHECK-DAG: [[BPADDR4:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX4]] // CHECK-DAG: [[BPADDR4:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX4]]
// CHECK-DAG: [[PADDR4:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX4]] // CHECK-DAG: [[PADDR4:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX4]]
// The names below are not necessarily consistent with the names used for the // The names below are not necessarily consistent with the names used for the
// addresses above as some are repeated. // addresses above as some are repeated.
// CHECK-DAG: store i[[SZ]] [[VLA0]], i[[SZ]]* [[CBPADDR0:%.+]], // CHECK-DAG: store i[[SZ]] [[VLA0]], i[[SZ]]* [[CBPADDR0:%.+]],
// CHECK-DAG: store i[[SZ]] [[VLA0]], i[[SZ]]* [[CPADDR0:%.+]], // CHECK-DAG: store i[[SZ]] [[VLA0]], i[[SZ]]* [[CPADDR0:%.+]],
// CHECK-DAG: [[CBPADDR0]] = bitcast i8** {{%[^,]+}} to i[[SZ]]* // CHECK-DAG: [[CBPADDR0]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// CHECK-DAG: [[CPADDR0]] = bitcast i8** {{%[^,]+}} to i[[SZ]]* // CHECK-DAG: [[CPADDR0]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// CHECK-DAG: store i[[SZ]] 2, i[[SZ]]* [[CBPADDR1:%.+]], // CHECK-DAG: store i[[SZ]] 2, i[[SZ]]* [[CBPADDR1:%.+]],
// CHECK-DAG: store i[[SZ]] 2, i[[SZ]]* [[CPADDR1:%.+]], // CHECK-DAG: store i[[SZ]] 2, i[[SZ]]* [[CPADDR1:%.+]],
// CHECK-DAG: [[CBPADDR1]] = bitcast i8** {{%[^,]+}} to i[[SZ]]* // CHECK-DAG: [[CBPADDR1]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// CHECK-DAG: [[CPADDR1]] = bitcast i8** {{%[^,]+}} to i[[SZ]]* // CHECK-DAG: [[CPADDR1]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// CHECK-DAG: store i[[SZ]] [[B_CVAL]], i[[SZ]]* [[CBPADDR2:%.+]], // CHECK-DAG: store i[[SZ]] [[B_CVAL]], i[[SZ]]* [[CBPADDR2:%.+]],
// CHECK-DAG: store i[[SZ]] [[B_CVAL]], i[[SZ]]* [[CPADDR2:%.+]], // CHECK-DAG: store i[[SZ]] [[B_CVAL]], i[[SZ]]* [[CPADDR2:%.+]],
// CHECK-DAG: [[CBPADDR2]] = bitcast i8** {{%[^,]+}} to i[[SZ]]* // CHECK-DAG: [[CBPADDR2]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// CHECK-DAG: [[CPADDR2]] = bitcast i8** {{%[^,]+}} to i[[SZ]]* // CHECK-DAG: [[CPADDR2]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// CHECK-DAG: store [[S1]]* %{{.+}}, [[S1]]** [[CBPADDR3:%.+]], // CHECK-DAG: store [[S1]]* %{{.+}}, [[S1]]** [[CBPADDR3:%.+]],
// CHECK-DAG: store [[S1]]* %{{.+}}, [[S1]]** [[CPADDR3:%.+]], // CHECK-DAG: store double* %{{.+}}, double** [[CPADDR3:%.+]],
// CHECK-DAG: [[CBPADDR3]] = bitcast i8** {{%[^,]+}} to [[S1]]** // CHECK-DAG: [[CBPADDR3]] = bitcast i8** {{%[^,]+}} to [[S1]]**
// CHECK-DAG: [[CPADDR3]] = bitcast i8** {{%[^,]+}} to [[S1]]** // CHECK-DAG: [[CPADDR3]] = bitcast i8** {{%[^,]+}} to double*
// CHECK-DAG: store i16* %{{.+}}, i16** [[CBPADDR4:%.+]], // CHECK-DAG: store i16* %{{.+}}, i16** [[CBPADDR4:%.+]],
// CHECK-DAG: store i16* %{{.+}}, i16** [[CPADDR4:%.+]], // CHECK-DAG: store i16* %{{.+}}, i16** [[CPADDR4:%.+]],
// CHECK-DAG: [[CBPADDR4]] = bitcast i8** {{%[^,]+}} to i16** // CHECK-DAG: [[CBPADDR4]] = bitcast i8** {{%[^,]+}} to i16**
// CHECK-DAG: [[CPADDR4]] = bitcast i8** {{%[^,]+}} to i16** // CHECK-DAG: [[CPADDR4]] = bitcast i8** {{%[^,]+}} to i16**
// CHECK-DAG: store i64 4, i64* {{%[^,]+}} // CHECK-DAG: store i64 4, i64* {{%[^,]+}}
// CHECK-DAG: store i64 [[CSIZE]], i64* {{%[^,]+}} // CHECK-DAG: store i64 [[CSIZE]], i64* {{%[^,]+}}
// CHECK-DAG: store i64 {{4|8}}, i64* {{%[^,]+}} // CHECK-DAG: store i64 {{4|8}}, i64* {{%[^,]+}}
// CHECK-DAG: store i64 {{4|8}}, i64* {{%[^,]+}} // CHECK-DAG: store i64 {{4|8}}, i64* {{%[^,]+}}
// CHECK-DAG: store i64 %{{.+}}, i64* {{%[^,]+}} // CHECK-DAG: store i64 {{4|8}}, i64* {{%[^,]+}}
// CHECK-NEXT: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0 // CHECK-NEXT: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0
// CHECK-NEXT: br i1 [[ERROR]], label %[[FAIL:[^,]+]], label %[[END:[^,]+]] // CHECK-NEXT: br i1 [[ERROR]], label %[[FAIL:[^,]+]], label %[[END:[^,]+]]
// CHECK: [[FAIL]] // CHECK: [[FAIL]]
// CHECK: call void [[HVT7:@.+]]({{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}) // CHECK: call void [[HVT7:@.+]]({{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}})
// CHECK-NEXT: br label %[[END]] // CHECK-NEXT: br label %[[END]]
// CHECK: [[END]] // CHECK: [[END]]
▲ Show 20 Lines • Show All 211 Lines • Show Last 20 Lines

clang/test/OpenMP/target_parallel_for_codegen.cpp

Show First 20 Lines • Show All 95 Lines • ▼ Show 20 Lines
// CHECK-DAG: [[MAPT2:@.+]] = private unnamed_addr constant [3 x i64] [i64 800, i64 800, i64 800] // CHECK-DAG: [[MAPT2:@.+]] = private unnamed_addr constant [3 x i64] [i64 800, i64 800, i64 800]
// CHECK-DAG: [[SIZET3:@.+]] = private unnamed_addr constant [2 x i64] [i64 4, i64 2] // CHECK-DAG: [[SIZET3:@.+]] = private unnamed_addr constant [2 x i64] [i64 4, i64 2]
// CHECK-DAG: [[MAPT3:@.+]] = private unnamed_addr constant [2 x i64] [i64 800, i64 800] // CHECK-DAG: [[MAPT3:@.+]] = private unnamed_addr constant [2 x i64] [i64 800, i64 800]
// CHECK-DAG: [[MAPT4:@.+]] = private unnamed_addr constant [10 x i64] [i64 800, i64 547, i64 800, i64 547, i64 547, i64 800, i64 800, i64 547, i64 547, i64 800] // CHECK-DAG: [[MAPT4:@.+]] = private unnamed_addr constant [10 x i64] [i64 800, i64 547, i64 800, i64 547, i64 547, i64 800, i64 800, i64 547, i64 547, i64 800]
// CHECK-DAG: [[SIZET5:@.+]] = private unnamed_addr constant [3 x i64] [i64 4, i64 2, i64 40] // CHECK-DAG: [[SIZET5:@.+]] = private unnamed_addr constant [3 x i64] [i64 4, i64 2, i64 40]
// CHECK-DAG: [[MAPT5:@.+]] = private unnamed_addr constant [3 x i64] [i64 800, i64 800, i64 547] // CHECK-DAG: [[MAPT5:@.+]] = private unnamed_addr constant [3 x i64] [i64 800, i64 800, i64 547]
// CHECK-DAG: [[SIZET6:@.+]] = private unnamed_addr constant [4 x i64] [i64 4, i64 2, i64 1, i64 40] // CHECK-DAG: [[SIZET6:@.+]] = private unnamed_addr constant [4 x i64] [i64 4, i64 2, i64 1, i64 40]
// CHECK-DAG: [[MAPT6:@.+]] = private unnamed_addr constant [4 x i64] [i64 800, i64 800, i64 800, i64 547] // CHECK-DAG: [[MAPT6:@.+]] = private unnamed_addr constant [4 x i64] [i64 800, i64 800, i64 800, i64 547]
// CHECK-DAG: [[MAPT7:@.+]] = private unnamed_addr constant [6 x i64] [i64 32, i64 281474976711171, i64 800, i64 800, i64 800, i64 547] // CHECK-DAG: [[MAPT7:@.+]] = private unnamed_addr constant [5 x i64] [i64 547, i64 800, i64 800, i64 800, i64 547]
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
▲ Show 20 Lines • Show All 516 Lines • ▼ Show 20 Lines
// CHECK: br i1 [[IF]], label %[[TRY:[^,]+]], label %[[FAIL:[^,]+]] // CHECK: br i1 [[IF]], label %[[TRY:[^,]+]], label %[[FAIL:[^,]+]]
// CHECK: [[TRY]] // CHECK: [[TRY]]
// We capture 2 VLA sizes in this target region // We capture 2 VLA sizes in this target region
// CHECK: [[CELEMSIZE2:%.+]] = mul nuw i[[SZ]] 2, [[VLA0:%.+]] // CHECK: [[CELEMSIZE2:%.+]] = mul nuw i[[SZ]] 2, [[VLA0:%.+]]
// CHECK-64: [[CSIZE:%.+]] = mul nuw i64 [[CELEMSIZE2]], 2 // CHECK-64: [[CSIZE:%.+]] = mul nuw i64 [[CELEMSIZE2]], 2
// CHECK-32: [[CSZSIZE:%.+]] = mul nuw i32 [[CELEMSIZE2]], 2 // CHECK-32: [[CSZSIZE:%.+]] = mul nuw i32 [[CELEMSIZE2]], 2
// CHECK-32: [[CSIZE:%.+]] = sext i32 [[CSZSIZE]] to i64 // CHECK-32: [[CSIZE:%.+]] = sext i32 [[CSZSIZE]] to i64
// CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 6, i8** [[BPR:%[^,]+]], i8** [[PR:%[^,]+]], i64* [[SR:%[^,]+]], i64* getelementptr inbounds ([6 x i64], [6 x i64]* [[MAPT7]], i32 0, i32 0), i8** null, i32 1, i32 0) // CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 5, i8** [[BPR:%[^,]+]], i8** [[PR:%[^,]+]], i64* [[SR:%[^,]+]], i64* getelementptr inbounds ([5 x i64], [5 x i64]* [[MAPT7]], i32 0, i32 0), i8** null, i32 1, i32 0)
// CHECK-DAG: [[BPR]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP:%.+]], i32 0, i32 0 // CHECK-DAG: [[BPR]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP:%.+]], i32 0, i32 0
// CHECK-DAG: [[PR]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P:%.+]], i32 0, i32 0 // CHECK-DAG: [[PR]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P:%.+]], i32 0, i32 0
// CHECK-DAG: [[SR]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S:%.+]], i32 0, i32 0 // CHECK-DAG: [[SR]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S:%.+]], i32 0, i32 0
// CHECK-DAG: [[SADDR0:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX0:[0-9]+]] // CHECK-DAG: [[SADDR0:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX0:[0-9]+]]
// CHECK-DAG: [[BPADDR0:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX0]] // CHECK-DAG: [[BPADDR0:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX0]]
// CHECK-DAG: [[PADDR0:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX0]] // CHECK-DAG: [[PADDR0:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX0]]
// CHECK-DAG: [[SADDR1:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX1:[0-9]+]] // CHECK-DAG: [[SADDR1:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX1:[0-9]+]]
// CHECK-DAG: [[BPADDR1:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX1]] // CHECK-DAG: [[BPADDR1:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX1]]
// CHECK-DAG: [[PADDR1:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX1]] // CHECK-DAG: [[PADDR1:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX1]]
// CHECK-DAG: [[SADDR2:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX2:[0-9]+]] // CHECK-DAG: [[SADDR2:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX2:[0-9]+]]
// CHECK-DAG: [[BPADDR2:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX2]] // CHECK-DAG: [[BPADDR2:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX2]]
// CHECK-DAG: [[PADDR2:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX2]] // CHECK-DAG: [[PADDR2:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX2]]
// CHECK-DAG: [[SADDR3:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX3:[0-9]+]] // CHECK-DAG: [[SADDR3:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX3:[0-9]+]]
// CHECK-DAG: [[BPADDR3:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX3]] // CHECK-DAG: [[BPADDR3:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX3]]
// CHECK-DAG: [[PADDR3:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX3]] // CHECK-DAG: [[PADDR3:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX3]]
// CHECK-DAG: [[SADDR4:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX4:[0-9]+]] // CHECK-DAG: [[SADDR4:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX4:[0-9]+]]
// CHECK-DAG: [[BPADDR4:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX4]] // CHECK-DAG: [[BPADDR4:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX4]]
// CHECK-DAG: [[PADDR4:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX4]] // CHECK-DAG: [[PADDR4:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX4]]
// CHECK-DAG: [[SADDR5:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX5:[0-9]+]]
// CHECK-DAG: [[BPADDR5:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX5]]
// CHECK-DAG: [[PADDR5:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX5]]
// The names below are not necessarily consistent with the names used for the // The names below are not necessarily consistent with the names used for the
// addresses above as some are repeated. // addresses above as some are repeated.
// CHECK-DAG: store i[[SZ]] [[VLA0]], i[[SZ]]* [[CBPADDR0:%.+]], // CHECK-DAG: store i[[SZ]] [[VLA0]], i[[SZ]]* [[CBPADDR0:%.+]],
// CHECK-DAG: store i[[SZ]] [[VLA0]], i[[SZ]]* [[CPADDR0:%.+]], // CHECK-DAG: store i[[SZ]] [[VLA0]], i[[SZ]]* [[CPADDR0:%.+]],
// CHECK-DAG: [[CBPADDR0]] = bitcast i8** {{%[^,]+}} to i[[SZ]]* // CHECK-DAG: [[CBPADDR0]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// CHECK-DAG: [[CPADDR0]] = bitcast i8** {{%[^,]+}} to i[[SZ]]* // CHECK-DAG: [[CPADDR0]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// CHECK-DAG: store i64 {{4|8}}, i64* {{%[^,]+}} // CHECK-DAG: store i64 {{4|8}}, i64* {{%[^,]+}}
Show All 11 Lines
// CHECK-DAG: store i64 4, i64* {{%[^,]+}} // CHECK-DAG: store i64 4, i64* {{%[^,]+}}
// CHECK-DAG: store [[S1]]* %{{.+}}, [[S1]]** [[CBPADDR3:%.+]], // CHECK-DAG: store [[S1]]* %{{.+}}, [[S1]]** [[CBPADDR3:%.+]],
// CHECK-DAG: store double* %{{.+}}, double** [[CPADDR3:%.+]], // CHECK-DAG: store double* %{{.+}}, double** [[CPADDR3:%.+]],
// CHECK-DAG: [[CBPADDR3]] = bitcast i8** {{%[^,]+}} to [[S1]]** // CHECK-DAG: [[CBPADDR3]] = bitcast i8** {{%[^,]+}} to [[S1]]**
// CHECK-DAG: [[CPADDR3]] = bitcast i8** {{%[^,]+}} to double** // CHECK-DAG: [[CPADDR3]] = bitcast i8** {{%[^,]+}} to double**
// CHECK-DAG: store i64 8, i64* {{%[^,]+}} // CHECK-DAG: store i64 8, i64* {{%[^,]+}}
// CHECK-DAG: store [[S1]]* %{{.+}}, [[S1]]** [[CBPADDR4:%.+]], // CHECK-DAG: store i16* %{{.+}}, i16** [[CBPADDR4:%.+]],
// CHECK-DAG: store double* %{{.+}}, double** [[CPADDR4:%.+]], // CHECK-DAG: store i16* %{{.+}}, i16** [[CPADDR4:%.+]],
// CHECK-DAG: [[CBPADDR4]] = bitcast i8** {{%[^,]+}} to [[S1]]** // CHECK-DAG: [[CBPADDR4]] = bitcast i8** {{%[^,]+}} to i16**
// CHECK-DAG: [[CPADDR4]] = bitcast i8** {{%[^,]+}} to double** // CHECK-DAG: [[CPADDR4]] = bitcast i8** {{%[^,]+}} to i16**
// CHECK-DAG: store i64 8, i64* {{%[^,]+}}
// CHECK-DAG: store i16* %{{.+}}, i16** [[CBPADDR5:%.+]],
// CHECK-DAG: store i16* %{{.+}}, i16** [[CPADDR5:%.+]],
// CHECK-DAG: [[CBPADDR5]] = bitcast i8** {{%[^,]+}} to i16**
// CHECK-DAG: [[CPADDR5]] = bitcast i8** {{%[^,]+}} to i16**
// CHECK-DAG: store i64 [[CSIZE]], i64* {{%[^,]+}} // CHECK-DAG: store i64 [[CSIZE]], i64* {{%[^,]+}}
// CHECK-NEXT: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0 // CHECK-NEXT: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0
// CHECK-NEXT: br i1 [[ERROR]], label %[[FAIL:[^,]+]], label %[[END:[^,]+]] // CHECK-NEXT: br i1 [[ERROR]], label %[[FAIL:[^,]+]], label %[[END:[^,]+]]
// CHECK: [[FAIL]] // CHECK: [[FAIL]]
// CHECK: call void [[HVT7:@.+]]({{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}) // CHECK: call void [[HVT7:@.+]]({{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}})
// CHECK-NEXT: br label %[[END]] // CHECK-NEXT: br label %[[END]]
▲ Show 20 Lines • Show All 212 Lines • Show Last 20 Lines

clang/test/OpenMP/target_parallel_for_simd_codegen.cpp

Show First 20 Lines • Show All 85 Lines • ▼ Show 20 Lines
// CHECK-DAG: [[MAPT2:@.+]] = private unnamed_addr constant [3 x i64] [i64 800, i64 800, i64 800] // CHECK-DAG: [[MAPT2:@.+]] = private unnamed_addr constant [3 x i64] [i64 800, i64 800, i64 800]
// CHECK-DAG: [[SIZET3:@.+]] = private unnamed_addr constant [2 x i64] [i64 4, i64 2] // CHECK-DAG: [[SIZET3:@.+]] = private unnamed_addr constant [2 x i64] [i64 4, i64 2]
// CHECK-DAG: [[MAPT3:@.+]] = private unnamed_addr constant [2 x i64] [i64 800, i64 800] // CHECK-DAG: [[MAPT3:@.+]] = private unnamed_addr constant [2 x i64] [i64 800, i64 800]
// CHECK-DAG: [[MAPT4:@.+]] = private unnamed_addr constant [10 x i64] [i64 800, i64 547, i64 800, i64 547, i64 547, i64 800, i64 800, i64 547, i64 547, i64 800] // CHECK-DAG: [[MAPT4:@.+]] = private unnamed_addr constant [10 x i64] [i64 800, i64 547, i64 800, i64 547, i64 547, i64 800, i64 800, i64 547, i64 547, i64 800]
// CHECK-DAG: [[SIZET5:@.+]] = private unnamed_addr constant [3 x i64] [i64 4, i64 2, i64 40] // CHECK-DAG: [[SIZET5:@.+]] = private unnamed_addr constant [3 x i64] [i64 4, i64 2, i64 40]
// CHECK-DAG: [[MAPT5:@.+]] = private unnamed_addr constant [3 x i64] [i64 800, i64 800, i64 547] // CHECK-DAG: [[MAPT5:@.+]] = private unnamed_addr constant [3 x i64] [i64 800, i64 800, i64 547]
// CHECK-DAG: [[SIZET6:@.+]] = private unnamed_addr constant [4 x i64] [i64 4, i64 2, i64 1, i64 40] // CHECK-DAG: [[SIZET6:@.+]] = private unnamed_addr constant [4 x i64] [i64 4, i64 2, i64 1, i64 40]
// CHECK-DAG: [[MAPT6:@.+]] = private unnamed_addr constant [4 x i64] [i64 800, i64 800, i64 800, i64 547] // CHECK-DAG: [[MAPT6:@.+]] = private unnamed_addr constant [4 x i64] [i64 800, i64 800, i64 800, i64 547]
// OMP45-DAG: [[MAPT7:@.+]] = private unnamed_addr constant [6 x i64] [i64 32, i64 281474976711171, i64 800, i64 800, i64 800, i64 547] // OMP45-DAG: [[MAPT7:@.+]] = private unnamed_addr constant [5 x i64] [i64 547, i64 800, i64 800, i64 800, i64 547]
// OMP50-DAG: [[MAPT7:@.+]] = private unnamed_addr constant [7 x i64] [i64 32, i64 281474976711171, i64 800, i64 800, i64 800, i64 547, i64 800] // OMP50-DAG: [[MAPT7:@.+]] = private unnamed_addr constant [6 x i64] [i64 547, i64 800, i64 800, i64 800, i64 547, i64 800]
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
▲ Show 20 Lines • Show All 499 Lines • ▼ Show 20 Lines
// CHECK: br i1 [[IF]], label %[[TRY:[^,]+]], label %[[FAIL:[^,]+]] // CHECK: br i1 [[IF]], label %[[TRY:[^,]+]], label %[[FAIL:[^,]+]]
// CHECK: [[TRY]] // CHECK: [[TRY]]
// We capture 2 VLA sizes in this target region // We capture 2 VLA sizes in this target region
// CHECK: [[CELEMSIZE2:%.+]] = mul nuw i[[SZ]] 2, [[VLA0:%.+]] // CHECK: [[CELEMSIZE2:%.+]] = mul nuw i[[SZ]] 2, [[VLA0:%.+]]
// CHECK-64: [[CSIZE:%.+]] = mul nuw i64 [[CELEMSIZE2]], 2 // CHECK-64: [[CSIZE:%.+]] = mul nuw i64 [[CELEMSIZE2]], 2
// CHECK-32: [[CSZSIZE:%.+]] = mul nuw i32 [[CELEMSIZE2]], 2 // CHECK-32: [[CSZSIZE:%.+]] = mul nuw i32 [[CELEMSIZE2]], 2
// CHECK-32: [[CSIZE:%.+]] = sext i32 [[CSZSIZE]] to i64 // CHECK-32: [[CSIZE:%.+]] = sext i32 [[CSZSIZE]] to i64
// OMP45-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 6, i8** [[BPR:%[^,]+]], i8** [[PR:%[^,]+]], i64* [[SR:%[^,]+]], i64* getelementptr inbounds ([6 x i64], [6 x i64]* [[MAPT7]], i32 0, i32 0), i8** null, i32 1, i32 0) // OMP45-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 5, i8** [[BPR:%[^,]+]], i8** [[PR:%[^,]+]], i64* [[SR:%[^,]+]], i64* getelementptr inbounds ([5 x i64], [5 x i64]* [[MAPT7]], i32 0, i32 0), i8** null, i32 1, i32 0)
// OMP50-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 7, i8** [[BPR:%[^,]+]], i8** [[PR:%[^,]+]], i64* [[SR:%[^,]+]], i64* getelementptr inbounds ([7 x i64], [7 x i64]* [[MAPT7]], i32 0, i32 0), i8** null, i32 1, i32 %{{.+}}) // OMP50-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 6, i8** [[BPR:%[^,]+]], i8** [[PR:%[^,]+]], i64* [[SR:%[^,]+]], i64* getelementptr inbounds ([6 x i64], [6 x i64]* [[MAPT7]], i32 0, i32 0), i8** null, i32 1, i32 %{{.+}})
// OMP45-DAG: [[BPR]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP:%.+]], i32 0, i32 0 // OMP45-DAG: [[BPR]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP:%.+]], i32 0, i32 0
// OMP45-DAG: [[PR]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P:%.+]], i32 0, i32 0 // OMP45-DAG: [[PR]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P:%.+]], i32 0, i32 0
// OMP45-DAG: [[SR]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S:%.+]], i32 0, i32 0 // OMP45-DAG: [[SR]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S:%.+]], i32 0, i32 0
// OMP45-DAG: [[SADDR0:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX0:[0-9]+]] // OMP45-DAG: [[SADDR0:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX0:[0-9]+]]
// OMP45-DAG: [[BPADDR0:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX0]] // OMP45-DAG: [[BPADDR0:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX0]]
// OMP45-DAG: [[PADDR0:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX0]] // OMP45-DAG: [[PADDR0:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX0]]
// OMP45-DAG: [[SADDR1:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX1:[0-9]+]] // OMP45-DAG: [[SADDR1:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX1:[0-9]+]]
// OMP45-DAG: [[BPADDR1:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX1]] // OMP45-DAG: [[BPADDR1:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX1]]
// OMP45-DAG: [[PADDR1:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX1]] // OMP45-DAG: [[PADDR1:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX1]]
// OMP45-DAG: [[SADDR2:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX2:[0-9]+]] // OMP45-DAG: [[SADDR2:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX2:[0-9]+]]
// OMP45-DAG: [[BPADDR2:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX2]] // OMP45-DAG: [[BPADDR2:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX2]]
// OMP45-DAG: [[PADDR2:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX2]] // OMP45-DAG: [[PADDR2:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX2]]
// OMP45-DAG: [[SADDR3:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX3:[0-9]+]] // OMP45-DAG: [[SADDR3:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX3:[0-9]+]]
// OMP45-DAG: [[BPADDR3:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX3]] // OMP45-DAG: [[BPADDR3:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX3]]
// OMP45-DAG: [[PADDR3:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX3]] // OMP45-DAG: [[PADDR3:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX3]]
// OMP45-DAG: [[SADDR4:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX4:[0-9]+]] // OMP45-DAG: [[SADDR4:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX4:[0-9]+]]
// OMP45-DAG: [[BPADDR4:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX4]] // OMP45-DAG: [[BPADDR4:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX4]]
// OMP45-DAG: [[PADDR4:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX4]] // OMP45-DAG: [[PADDR4:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX4]]
// OMP45-DAG: [[SADDR5:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX5:[0-9]+]] // OMP50-DAG: [[BPR]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP:%.+]], i32 0, i32 0
// OMP45-DAG: [[BPADDR5:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX5]] // OMP50-DAG: [[PR]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P:%.+]], i32 0, i32 0
// OMP45-DAG: [[PADDR5:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX5]] // OMP50-DAG: [[SR]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S:%.+]], i32 0, i32 0
// OMP50-DAG: [[BPR]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[BP:%.+]], i32 0, i32 0 // OMP50-DAG: [[SADDR0:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX0:[0-9]+]]
// OMP50-DAG: [[PR]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[P:%.+]], i32 0, i32 0 // OMP50-DAG: [[BPADDR0:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX0]]
// OMP50-DAG: [[SR]] = getelementptr inbounds [7 x i64], [7 x i64]* [[S:%.+]], i32 0, i32 0 // OMP50-DAG: [[PADDR0:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX0]]
// OMP50-DAG: [[SADDR0:%.+]] = getelementptr inbounds [7 x i64], [7 x i64]* [[S]], i32 [[IDX0:[0-9]+]] // OMP50-DAG: [[SADDR1:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX1:[0-9]+]]
// OMP50-DAG: [[BPADDR0:%.+]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[BP]], i32 [[IDX0]] // OMP50-DAG: [[BPADDR1:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX1]]
// OMP50-DAG: [[PADDR0:%.+]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[P]], i32 [[IDX0]] // OMP50-DAG: [[PADDR1:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX1]]
// OMP50-DAG: [[SADDR1:%.+]] = getelementptr inbounds [7 x i64], [7 x i64]* [[S]], i32 [[IDX1:[0-9]+]] // OMP50-DAG: [[SADDR2:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX2:[0-9]+]]
// OMP50-DAG: [[BPADDR1:%.+]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[BP]], i32 [[IDX1]] // OMP50-DAG: [[BPADDR2:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX2]]
// OMP50-DAG: [[PADDR1:%.+]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[P]], i32 [[IDX1]] // OMP50-DAG: [[PADDR2:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX2]]
// OMP50-DAG: [[SADDR2:%.+]] = getelementptr inbounds [7 x i64], [7 x i64]* [[S]], i32 [[IDX2:[0-9]+]] // OMP50-DAG: [[SADDR3:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX3:[0-9]+]]
// OMP50-DAG: [[BPADDR2:%.+]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[BP]], i32 [[IDX2]] // OMP50-DAG: [[BPADDR3:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX3]]
// OMP50-DAG: [[PADDR2:%.+]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[P]], i32 [[IDX2]] // OMP50-DAG: [[PADDR3:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX3]]
// OMP50-DAG: [[SADDR3:%.+]] = getelementptr inbounds [7 x i64], [7 x i64]* [[S]], i32 [[IDX3:[0-9]+]] // OMP50-DAG: [[SADDR4:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX4:[0-9]+]]
// OMP50-DAG: [[BPADDR3:%.+]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[BP]], i32 [[IDX3]] // OMP50-DAG: [[BPADDR4:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX4]]
// OMP50-DAG: [[PADDR3:%.+]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[P]], i32 [[IDX3]] // OMP50-DAG: [[PADDR4:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX4]]
// OMP50-DAG: [[SADDR4:%.+]] = getelementptr inbounds [7 x i64], [7 x i64]* [[S]], i32 [[IDX4:[0-9]+]] // OMP50-DAG: [[SADDR5:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX5:[0-9]+]]
// OMP50-DAG: [[BPADDR4:%.+]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[BP]], i32 [[IDX4]] // OMP50-DAG: [[BPADDR5:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX5]]
// OMP50-DAG: [[PADDR4:%.+]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[P]], i32 [[IDX4]] // OMP50-DAG: [[PADDR5:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX5]]
// OMP50-DAG: [[SADDR5:%.+]] = getelementptr inbounds [7 x i64], [7 x i64]* [[S]], i32 [[IDX5:[0-9]+]]
// OMP50-DAG: [[BPADDR5:%.+]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[BP]], i32 [[IDX5]]
// OMP50-DAG: [[PADDR5:%.+]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[P]], i32 [[IDX5]]
// OMP50-DAG: [[SADDR6:%.+]] = getelementptr inbounds [7 x i64], [7 x i64]* [[S]], i32 [[IDX6:[0-9]+]]
// OMP50-DAG: [[BPADDR6:%.+]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[BP]], i32 [[IDX6]]
// OMP50-DAG: [[PADDR6:%.+]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[P]], i32 [[IDX6]]
// The names below are not necessarily consistent with the names used for the // The names below are not necessarily consistent with the names used for the
// addresses above as some are repeated. // addresses above as some are repeated.
// CHECK-DAG: store i[[SZ]] [[VLA0]], i[[SZ]]* [[CBPADDR0:%.+]], // CHECK-DAG: store i[[SZ]] [[VLA0]], i[[SZ]]* [[CBPADDR0:%.+]],
// CHECK-DAG: store i[[SZ]] [[VLA0]], i[[SZ]]* [[CPADDR0:%.+]], // CHECK-DAG: store i[[SZ]] [[VLA0]], i[[SZ]]* [[CPADDR0:%.+]],
// CHECK-DAG: [[CBPADDR0]] = bitcast i8** {{%[^,]+}} to i[[SZ]]* // CHECK-DAG: [[CBPADDR0]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// CHECK-DAG: [[CPADDR0]] = bitcast i8** {{%[^,]+}} to i[[SZ]]* // CHECK-DAG: [[CPADDR0]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// CHECK-DAG: store i64 {{4|8}}, i64* {{%[^,]+}} // CHECK-DAG: store i64 {{4|8}}, i64* {{%[^,]+}}
Show All 11 Lines
// CHECK-DAG: store i64 4, i64* {{%[^,]+}} // CHECK-DAG: store i64 4, i64* {{%[^,]+}}
// CHECK-DAG: store [[S1]]* %{{.+}}, [[S1]]** [[CBPADDR3:%.+]], // CHECK-DAG: store [[S1]]* %{{.+}}, [[S1]]** [[CBPADDR3:%.+]],
// CHECK-DAG: store double* %{{.+}}, double** [[CPADDR3:%.+]], // CHECK-DAG: store double* %{{.+}}, double** [[CPADDR3:%.+]],
// CHECK-DAG: [[CBPADDR3]] = bitcast i8** {{%[^,]+}} to [[S1]]** // CHECK-DAG: [[CBPADDR3]] = bitcast i8** {{%[^,]+}} to [[S1]]**
// CHECK-DAG: [[CPADDR3]] = bitcast i8** {{%[^,]+}} to double** // CHECK-DAG: [[CPADDR3]] = bitcast i8** {{%[^,]+}} to double**
// CHECK-DAG: store i64 8, i64* {{%[^,]+}} // CHECK-DAG: store i64 8, i64* {{%[^,]+}}
// CHECK-DAG: store [[S1]]* %{{.+}}, [[S1]]** [[CBPADDR4:%.+]], // CHECK-DAG: store i16* %{{.+}}, i16** [[CBPADDR4:%.+]],
// CHECK-DAG: store double* %{{.+}}, double** [[CPADDR4:%.+]], // CHECK-DAG: store i16* %{{.+}}, i16** [[CPADDR4:%.+]],
// CHECK-DAG: [[CBPADDR4]] = bitcast i8** {{%[^,]+}} to [[S1]]** // CHECK-DAG: [[CBPADDR4]] = bitcast i8** {{%[^,]+}} to i16**
// CHECK-DAG: [[CPADDR4]] = bitcast i8** {{%[^,]+}} to double** // CHECK-DAG: [[CPADDR4]] = bitcast i8** {{%[^,]+}} to i16**
// CHECK-DAG: store i64 8, i64* {{%[^,]+}}
// CHECK-DAG: store i16* %{{.+}}, i16** [[CBPADDR5:%.+]],
// CHECK-DAG: store i16* %{{.+}}, i16** [[CPADDR5:%.+]],
// CHECK-DAG: [[CBPADDR5]] = bitcast i8** {{%[^,]+}} to i16**
// CHECK-DAG: [[CPADDR5]] = bitcast i8** {{%[^,]+}} to i16**
// CHECK-DAG: store i64 [[CSIZE]], i64* {{%[^,]+}} // CHECK-DAG: store i64 [[CSIZE]], i64* {{%[^,]+}}
// OMP50-DAG: store i[[SZ]] [[SIMD_COND]], i[[SZ]]* [[CBPADDR6:%.+]], // OMP50-DAG: store i[[SZ]] [[SIMD_COND]], i[[SZ]]* [[CBPADDR5:%.+]],
// OMP50-DAG: store i[[SZ]] [[SIMD_COND]], i[[SZ]]* [[CPADDR6:%.+]], // OMP50-DAG: store i[[SZ]] [[SIMD_COND]], i[[SZ]]* [[CPADDR5:%.+]],
// OMP50-DAG: [[CBPADDR6]] = bitcast i8** {{%[^,]+}} to i[[SZ]]* // OMP50-DAG: [[CBPADDR5]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// OMP50-DAG: [[CPADDR6]] = bitcast i8** {{%[^,]+}} to i[[SZ]]* // OMP50-DAG: [[CPADDR5]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// OMP50-DAG: store i64 1, i64* {{%[^,]+}} // OMP50-DAG: store i64 1, i64* {{%[^,]+}}
// CHECK-NEXT: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0 // CHECK-NEXT: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0
// CHECK-NEXT: br i1 [[ERROR]], label %[[FAIL:[^,]+]], label %[[END:[^,]+]] // CHECK-NEXT: br i1 [[ERROR]], label %[[FAIL:[^,]+]], label %[[END:[^,]+]]
// CHECK: [[FAIL]] // CHECK: [[FAIL]]
// OMP45: call void [[HVT7:@.+]]({{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}) // OMP45: call void [[HVT7:@.+]]({{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}})
// OMP50: call void [[HVT7:@.+]]({{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}) // OMP50: call void [[HVT7:@.+]]({{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}})
▲ Show 20 Lines • Show All 225 Lines • Show Last 20 Lines

clang/test/OpenMP/target_parallel_if_codegen.cpp

Show First 20 Lines • Show All 180 Lines • ▼ Show 20 Lines
// CHECK: store i8 [[FB]], i8* [[CAPE_ADDR:%.+]], align // CHECK: store i8 [[FB]], i8* [[CAPE_ADDR:%.+]], align
// CHECK: [[CAPE:%.+]] = load i8, i8* [[CAPE_ADDR]], align // CHECK: [[CAPE:%.+]] = load i8, i8* [[CAPE_ADDR]], align
// CHECK: [[TB:%.+]] = trunc i8 [[CAPE]] to i1 // CHECK: [[TB:%.+]] = trunc i8 [[CAPE]] to i1
// CHECK: [[CONV:%.+]] = bitcast i[[SZ]]* [[CAPEC_ADDR:%.+]] to i8* // CHECK: [[CONV:%.+]] = bitcast i[[SZ]]* [[CAPEC_ADDR:%.+]] to i8*
// CHECK: [[FB:%.+]] = zext i1 [[TB]] to i8 // CHECK: [[FB:%.+]] = zext i1 [[TB]] to i8
// CHECK: store i8 [[FB]], i8* [[CONV]], align // CHECK: store i8 [[FB]], i8* [[CONV]], align
// CHECK: [[ARG:%.+]] = load i[[SZ]], i[[SZ]]* [[CAPEC_ADDR]], align // CHECK: [[ARG:%.+]] = load i[[SZ]], i[[SZ]]* [[CAPEC_ADDR]], align
// //
// CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 4, {{.*}}, i8** null, i32 1, i32 [[NT:%.+]]) // CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 3, {{.*}}, i8** null, i32 1, i32 [[NT:%.+]])
// CHECK-DAG: [[NT]] = select i1 %{{.+}}, i32 0, i32 1 // CHECK-DAG: [[NT]] = select i1 %{{.+}}, i32 0, i32 1
// CHECK: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0 // CHECK: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0
// CHECK: br i1 [[ERROR]], label %[[FAIL:.+]], label %[[END:[^,]+]] // CHECK: br i1 [[ERROR]], label %[[FAIL:.+]], label %[[END:[^,]+]]
// //
// CHECK: [[FAIL]] // CHECK: [[FAIL]]
// CHECK: call void [[HVT1:@.+]]([[S1]]* {{%.+}}, i[[SZ]] {{%.+}}, i[[SZ]] [[ARG]]) // CHECK: call void [[HVT1:@.+]]([[S1]]* {{%.+}}, i[[SZ]] {{%.+}}, i[[SZ]] [[ARG]])
// CHECK: br label {{%?}}[[END]] // CHECK: br label {{%?}}[[END]]
// CHECK: [[END]] // CHECK: [[END]]
Show All 10 Lines
// CHECK: [[FB:%.+]] = zext i1 [[TB]] to i8 // CHECK: [[FB:%.+]] = zext i1 [[TB]] to i8
// CHECK: store i8 [[FB]], i8* [[CONV]], align // CHECK: store i8 [[FB]], i8* [[CONV]], align
// CHECK: [[ARG:%.+]] = load i[[SZ]], i[[SZ]]* [[CAPEC_ADDR]], align // CHECK: [[ARG:%.+]] = load i[[SZ]], i[[SZ]]* [[CAPEC_ADDR]], align
// CHECK: [[NV:%.+]] = load i32, i32* [[N_ADDR]], align // CHECK: [[NV:%.+]] = load i32, i32* [[N_ADDR]], align
// CHECK: [[CMP:%.+]] = icmp sgt i32 [[NV]], 4 // CHECK: [[CMP:%.+]] = icmp sgt i32 [[NV]], 4
// CHECK: br i1 [[CMP]], label {{%?}}[[IF_THEN:.+]], label {{%?}}[[IF_ELSE:.+]] // CHECK: br i1 [[CMP]], label {{%?}}[[IF_THEN:.+]], label {{%?}}[[IF_ELSE:.+]]
// //
// CHECK: [[IF_THEN]] // CHECK: [[IF_THEN]]
// CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 3, {{.*}}, i8** null, i32 1, i32 [[NT:%.+]]) // CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 2, {{.*}}, i8** null, i32 1, i32 [[NT:%.+]])
// CHECK-DAG: [[NT]] = select i1 %{{.+}}, i32 0, i32 1 // CHECK-DAG: [[NT]] = select i1 %{{.+}}, i32 0, i32 1
// CHECK: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0 // CHECK: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0
// CHECK-NEXT: br i1 [[ERROR]], label %[[FAIL:.+]], label %[[END:[^,]+]] // CHECK-NEXT: br i1 [[ERROR]], label %[[FAIL:.+]], label %[[END:[^,]+]]
// CHECK: [[FAIL]] // CHECK: [[FAIL]]
// CHECK: call void [[HVT2:@.+]]([[S1]]* {{%.+}}, i[[SZ]] [[ARG]]) // CHECK: call void [[HVT2:@.+]]([[S1]]* {{%.+}}, i[[SZ]] [[ARG]])
// CHECK-NEXT: br label %[[END]] // CHECK-NEXT: br label %[[END]]
// CHECK: [[END]] // CHECK: [[END]]
// CHECK-NEXT: br label %[[IFEND:.+]] // CHECK-NEXT: br label %[[IFEND:.+]]
▲ Show 20 Lines • Show All 207 Lines • Show Last 20 Lines

clang/test/OpenMP/target_parallel_num_threads_codegen.cpp

Show First 20 Lines • Show All 183 Lines • ▼ Show 20 Lines
// CHECK: store i32 [[SUB]], i32* [[CAPE_ADDR:%.+]], align // CHECK: store i32 [[SUB]], i32* [[CAPE_ADDR:%.+]], align
// CHECK: [[CEV:%.+]] = load i32, i32* [[CAPE_ADDR]], align // CHECK: [[CEV:%.+]] = load i32, i32* [[CAPE_ADDR]], align
// CHECK-64: [[CONV:%.+]] = bitcast i[[SZ]]* [[CAPEC_ADDR:%.+]] to i32* // CHECK-64: [[CONV:%.+]] = bitcast i[[SZ]]* [[CAPEC_ADDR:%.+]] to i32*
// CHECK-64: store i32 [[CEV]], i32* [[CONV]], align // CHECK-64: store i32 [[CEV]], i32* [[CONV]], align
// CHECK-32: store i32 [[CEV]], i32* [[CAPEC_ADDR:%.+]], align // CHECK-32: store i32 [[CEV]], i32* [[CAPEC_ADDR:%.+]], align
// CHECK: [[ARG:%.+]] = load i[[SZ]], i[[SZ]]* [[CAPEC_ADDR]], align // CHECK: [[ARG:%.+]] = load i[[SZ]], i[[SZ]]* [[CAPEC_ADDR]], align
// CHECK: [[THREADS:%.+]] = load i32, i32* [[CAPE_ADDR]], align // CHECK: [[THREADS:%.+]] = load i32, i32* [[CAPE_ADDR]], align
// //
// CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 4, {{.*}}, i8** null, i32 1, i32 [[THREADS]]) // CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 3, {{.*}}, i8** null, i32 1, i32 [[THREADS]])
// CHECK: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0 // CHECK: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0
// CHECK: br i1 [[ERROR]], label %[[FAIL:.+]], label %[[END:[^,]+]] // CHECK: br i1 [[ERROR]], label %[[FAIL:.+]], label %[[END:[^,]+]]
// //
// CHECK: [[FAIL]] // CHECK: [[FAIL]]
// CHECK: call void [[HVT1:@.+]]([[S1]]* {{%.+}}, i[[SZ]] {{%.+}}, i[[SZ]] [[ARG]]) // CHECK: call void [[HVT1:@.+]]([[S1]]* {{%.+}}, i[[SZ]] {{%.+}}, i[[SZ]] [[ARG]])
// CHECK: br label {{%?}}[[END]] // CHECK: br label {{%?}}[[END]]
// CHECK: [[END]] // CHECK: [[END]]
// //
// //
// //
// CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 2, {{.+}}, i8** null, i32 1, i32 1024) // CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 1, {{.+}}, i8** null, i32 1, i32 1024)
// CHECK: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0 // CHECK: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0
// CHECK: br i1 [[ERROR]], label %[[FAIL:.+]], label %[[END:[^,]+]] // CHECK: br i1 [[ERROR]], label %[[FAIL:.+]], label %[[END:[^,]+]]
// //
// CHECK: [[FAIL]] // CHECK: [[FAIL]]
// CHECK: call void [[HVT2:@.+]]([[S1]]* {{[^,]+}}) // CHECK: call void [[HVT2:@.+]]([[S1]]* {{[^,]+}})
// CHECK: br label {{%?}}[[END]] // CHECK: br label {{%?}}[[END]]
// CHECK: [[END]] // CHECK: [[END]]
// //
▲ Show 20 Lines • Show All 165 Lines • Show Last 20 Lines

clang/test/OpenMP/target_teams_codegen.cpp

Show First 20 Lines • Show All 95 Lines • ▼ Show 20 Lines
// CHECK-DAG: [[MAPT2:@.+]] = private unnamed_addr constant [1 x i64] [i64 800] // CHECK-DAG: [[MAPT2:@.+]] = private unnamed_addr constant [1 x i64] [i64 800]
// CHECK-DAG: [[SIZET3:@.+]] = private unnamed_addr constant [2 x i64] [i64 4, i64 2] // CHECK-DAG: [[SIZET3:@.+]] = private unnamed_addr constant [2 x i64] [i64 4, i64 2]
// CHECK-DAG: [[MAPT3:@.+]] = private unnamed_addr constant [2 x i64] [i64 800, i64 800] // CHECK-DAG: [[MAPT3:@.+]] = private unnamed_addr constant [2 x i64] [i64 800, i64 800]
// CHECK-DAG: [[MAPT4:@.+]] = private unnamed_addr constant [9 x i64] [i64 800, i64 547, i64 800, i64 547, i64 547, i64 800, i64 800, i64 547, i64 547] // CHECK-DAG: [[MAPT4:@.+]] = private unnamed_addr constant [9 x i64] [i64 800, i64 547, i64 800, i64 547, i64 547, i64 800, i64 800, i64 547, i64 547]
// CHECK-DAG: [[SIZET5:@.+]] = private unnamed_addr constant [3 x i64] [i64 4, i64 2, i64 40] // CHECK-DAG: [[SIZET5:@.+]] = private unnamed_addr constant [3 x i64] [i64 4, i64 2, i64 40]
// CHECK-DAG: [[MAPT5:@.+]] = private unnamed_addr constant [3 x i64] [i64 800, i64 800, i64 547] // CHECK-DAG: [[MAPT5:@.+]] = private unnamed_addr constant [3 x i64] [i64 800, i64 800, i64 547]
// CHECK-DAG: [[SIZET6:@.+]] = private unnamed_addr constant [4 x i64] [i64 4, i64 2, i64 1, i64 40] // CHECK-DAG: [[SIZET6:@.+]] = private unnamed_addr constant [4 x i64] [i64 4, i64 2, i64 1, i64 40]
// CHECK-DAG: [[MAPT6:@.+]] = private unnamed_addr constant [4 x i64] [i64 800, i64 800, i64 800, i64 547] // CHECK-DAG: [[MAPT6:@.+]] = private unnamed_addr constant [4 x i64] [i64 800, i64 800, i64 800, i64 547]
// CHECK-DAG: [[MAPT7:@.+]] = private unnamed_addr constant [6 x i64] [i64 32, i64 281474976711171, i64 800, i64 800, i64 800, i64 547] // CHECK-DAG: [[MAPT7:@.+]] = private unnamed_addr constant [5 x i64] [i64 547, i64 800, i64 800, i64 800, i64 547]
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
▲ Show 20 Lines • Show All 511 Lines • ▼ Show 20 Lines
// CHECK: br i1 [[IF]], label %[[TRY:[^,]+]], label %[[FAIL:[^,]+]] // CHECK: br i1 [[IF]], label %[[TRY:[^,]+]], label %[[FAIL:[^,]+]]
// CHECK: [[TRY]] // CHECK: [[TRY]]
// We capture 2 VLA sizes in this target region // We capture 2 VLA sizes in this target region
// CHECK: [[CELEMSIZE2:%.+]] = mul nuw i[[SZ]] 2, [[VLA0:%.+]] // CHECK: [[CELEMSIZE2:%.+]] = mul nuw i[[SZ]] 2, [[VLA0:%.+]]
// CHECK-64: [[CSIZE:%.+]] = mul nuw i64 [[CELEMSIZE2]], 2 // CHECK-64: [[CSIZE:%.+]] = mul nuw i64 [[CELEMSIZE2]], 2
// CHECK-32: [[CSZSIZE:%.+]] = mul nuw i32 [[CELEMSIZE2]], 2 // CHECK-32: [[CSZSIZE:%.+]] = mul nuw i32 [[CELEMSIZE2]], 2
// CHECK-32: [[CSIZE:%.+]] = sext i32 [[CSZSIZE]] to i64 // CHECK-32: [[CSIZE:%.+]] = sext i32 [[CSZSIZE]] to i64
// CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 6, i8** [[BPR:%[^,]+]], i8** [[PR:%[^,]+]], i64* [[SR:%[^,]+]], i64* getelementptr inbounds ([6 x i64], [6 x i64]* [[MAPT7]], i32 0, i32 0), i8** null, i32 0, i32 0) // CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 5, i8** [[BPR:%[^,]+]], i8** [[PR:%[^,]+]], i64* [[SR:%[^,]+]], i64* getelementptr inbounds ([5 x i64], [5 x i64]* [[MAPT7]], i32 0, i32 0), i8** null, i32 0, i32 0)
// CHECK-DAG: [[BPR]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP:%.+]], i32 0, i32 0 // CHECK-DAG: [[BPR]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP:%.+]], i32 0, i32 0
// CHECK-DAG: [[PR]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P:%.+]], i32 0, i32 0 // CHECK-DAG: [[PR]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P:%.+]], i32 0, i32 0
// CHECK-DAG: [[SR]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S:%.+]], i32 0, i32 0 // CHECK-DAG: [[SR]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S:%.+]], i32 0, i32 0
// CHECK-DAG: [[SADDR0:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX0:[0-9]+]] // CHECK-DAG: [[SADDR0:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX0:[0-9]+]]
// CHECK-DAG: [[BPADDR0:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX0]] // CHECK-DAG: [[BPADDR0:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX0]]
// CHECK-DAG: [[PADDR0:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX0]] // CHECK-DAG: [[PADDR0:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX0]]
// CHECK-DAG: [[SADDR1:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX1:[0-9]+]] // CHECK-DAG: [[SADDR1:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX1:[0-9]+]]
// CHECK-DAG: [[BPADDR1:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX1]] // CHECK-DAG: [[BPADDR1:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX1]]
// CHECK-DAG: [[PADDR1:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX1]] // CHECK-DAG: [[PADDR1:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX1]]
// CHECK-DAG: [[SADDR2:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX2:[0-9]+]] // CHECK-DAG: [[SADDR2:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX2:[0-9]+]]
// CHECK-DAG: [[BPADDR2:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX2]] // CHECK-DAG: [[BPADDR2:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX2]]
// CHECK-DAG: [[PADDR2:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX2]] // CHECK-DAG: [[PADDR2:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX2]]
// CHECK-DAG: [[SADDR3:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX3:[0-9]+]] // CHECK-DAG: [[SADDR3:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX3:[0-9]+]]
// CHECK-DAG: [[BPADDR3:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX3]] // CHECK-DAG: [[BPADDR3:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX3]]
// CHECK-DAG: [[PADDR3:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX3]] // CHECK-DAG: [[PADDR3:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX3]]
// CHECK-DAG: [[SADDR4:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX4:[0-9]+]] // CHECK-DAG: [[SADDR4:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX4:[0-9]+]]
// CHECK-DAG: [[BPADDR4:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX4]] // CHECK-DAG: [[BPADDR4:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX4]]
// CHECK-DAG: [[PADDR4:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX4]] // CHECK-DAG: [[PADDR4:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX4]]
// The names below are not necessarily consistent with the names used for the // The names below are not necessarily consistent with the names used for the
// addresses above as some are repeated. // addresses above as some are repeated.
// CHECK-DAG: store i[[SZ]] [[VLA0]], i[[SZ]]* [[CBPADDR0:%.+]], // CHECK-DAG: store i[[SZ]] [[VLA0]], i[[SZ]]* [[CBPADDR0:%.+]],
// CHECK-DAG: store i[[SZ]] [[VLA0]], i[[SZ]]* [[CPADDR0:%.+]], // CHECK-DAG: store i[[SZ]] [[VLA0]], i[[SZ]]* [[CPADDR0:%.+]],
// CHECK-DAG: [[CBPADDR0]] = bitcast i8** {{%[^,]+}} to i[[SZ]]* // CHECK-DAG: [[CBPADDR0]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// CHECK-DAG: [[CPADDR0]] = bitcast i8** {{%[^,]+}} to i[[SZ]]* // CHECK-DAG: [[CPADDR0]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// CHECK-DAG: store i64 {{4|8}}, i64* {{%[^,]+}} // CHECK-DAG: store i64 {{4|8}}, i64* {{%[^,]+}}
// CHECK-DAG: store i[[SZ]] 2, i[[SZ]]* [[CBPADDR1:%.+]], // CHECK-DAG: store i[[SZ]] 2, i[[SZ]]* [[CBPADDR1:%.+]],
// CHECK-DAG: store i[[SZ]] 2, i[[SZ]]* [[CPADDR1:%.+]], // CHECK-DAG: store i[[SZ]] 2, i[[SZ]]* [[CPADDR1:%.+]],
// CHECK-DAG: [[CBPADDR1]] = bitcast i8** {{%[^,]+}} to i[[SZ]]* // CHECK-DAG: [[CBPADDR1]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// CHECK-DAG: [[CPADDR1]] = bitcast i8** {{%[^,]+}} to i[[SZ]]* // CHECK-DAG: [[CPADDR1]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// CHECK-DAG: store i64 {{4|8}}, i64* {{%[^,]+}} // CHECK-DAG: store i64 {{4|8}}, i64* {{%[^,]+}}
// CHECK-DAG: store i[[SZ]] [[B_CVAL]], i[[SZ]]* [[CBPADDR2:%.+]], // CHECK-DAG: store i[[SZ]] [[B_CVAL]], i[[SZ]]* [[CBPADDR2:%.+]],
// CHECK-DAG: store i[[SZ]] [[B_CVAL]], i[[SZ]]* [[CPADDR2:%.+]], // CHECK-DAG: store i[[SZ]] [[B_CVAL]], i[[SZ]]* [[CPADDR2:%.+]],
// CHECK-DAG: [[CBPADDR2]] = bitcast i8** {{%[^,]+}} to i[[SZ]]* // CHECK-DAG: [[CBPADDR2]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// CHECK-DAG: [[CPADDR2]] = bitcast i8** {{%[^,]+}} to i[[SZ]]* // CHECK-DAG: [[CPADDR2]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// CHECK-DAG: store i64 4, i64* {{%[^,]+}} // CHECK-DAG: store i64 4, i64* {{%[^,]+}}
// CHECK-DAG: store [[S1]]* %{{.+}}, [[S1]]** [[CBPADDR3:%.+]], // CHECK-DAG: store [[S1]]* %{{.+}}, [[S1]]** [[CBPADDR3:%.+]],
// CHECK-DAG: store [[S1]]* %{{.+}}, [[S1]]** [[CPADDR3:%.+]], // CHECK-DAG: store double* %{{.+}}, double** [[CPADDR3:%.+]],
// CHECK-DAG: [[CBPADDR3]] = bitcast i8** {{%[^,]+}} to [[S1]]** // CHECK-DAG: [[CBPADDR3]] = bitcast i8** {{%[^,]+}} to [[S1]]**
// CHECK-DAG: [[CPADDR3]] = bitcast i8** {{%[^,]+}} to [[S1]]** // CHECK-DAG: [[CPADDR3]] = bitcast i8** {{%[^,]+}} to double**
// CHECK-DAG: store i64 8, i64* {{%[^,]+}} // CHECK-DAG: store i64 8, i64* {{%[^,]+}}
// CHECK-DAG: store [[S1]]* %{{.+}}, [[S1]]** [[CBPADDR4:%.+]], // CHECK-DAG: store i16* %{{.+}}, i16** [[CBPADDR4:%.+]],
// CHECK-DAG: store [[S1]]* %{{.+}}, [[S1]]** [[CPADDR4:%.+]], // CHECK-DAG: store i16* %{{.+}}, i16** [[CPADDR4:%.+]],
// CHECK-DAG: [[CBPADDR4]] = bitcast i8** {{%[^,]+}} to [[S1]]** // CHECK-DAG: [[CBPADDR4]] = bitcast i8** {{%[^,]+}} to i16**
// CHECK-DAG: [[CPADDR4]] = bitcast i8** {{%[^,]+}} to [[S1]]** // CHECK-DAG: [[CPADDR4]] = bitcast i8** {{%[^,]+}} to i16**
// CHECK-DAG: store i64 8, i64* {{%[^,]+}}
// CHECK-DAG: store i16* %{{.+}}, i16** [[CBPADDR5:%.+]],
// CHECK-DAG: store i16* %{{.+}}, i16** [[CPADDR5:%.+]],
// CHECK-DAG: [[CBPADDR5]] = bitcast i8** {{%[^,]+}} to i16**
// CHECK-DAG: [[CPADDR5]] = bitcast i8** {{%[^,]+}} to i16**
// CHECK-DAG: store i64 [[CSIZE]], i64* {{%[^,]+}} // CHECK-DAG: store i64 [[CSIZE]], i64* {{%[^,]+}}
// CHECK-NEXT: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0 // CHECK-NEXT: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0
// CHECK-NEXT: br i1 [[ERROR]], label %[[FAIL:[^,]+]], label %[[END:[^,]+]] // CHECK-NEXT: br i1 [[ERROR]], label %[[FAIL:[^,]+]], label %[[END:[^,]+]]
// CHECK: [[FAIL]] // CHECK: [[FAIL]]
// CHECK: call void [[HVT7:@.+]]({{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}) // CHECK: call void [[HVT7:@.+]]({{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}})
// CHECK-NEXT: br label %[[END]] // CHECK-NEXT: br label %[[END]]
▲ Show 20 Lines • Show All 210 Lines • Show Last 20 Lines

clang/test/OpenMP/target_teams_distribute_codegen.cpp

Show First 20 Lines • Show All 93 Lines • ▼ Show 20 Lines
// CHECK-DAG: [[SIZET:@.+]] = private unnamed_addr constant [3 x i64] [i64 2, i64 4, i64 4] // CHECK-DAG: [[SIZET:@.+]] = private unnamed_addr constant [3 x i64] [i64 2, i64 4, i64 4]
// CHECK-DAG: [[MAPT:@.+]] = private unnamed_addr constant [3 x i64] [i64 288, i64 800, i64 800] // CHECK-DAG: [[MAPT:@.+]] = private unnamed_addr constant [3 x i64] [i64 288, i64 800, i64 800]
// CHECK-DAG: [[SIZET2:@.+]] = private unnamed_addr constant [1 x i64] [i64 2] // CHECK-DAG: [[SIZET2:@.+]] = private unnamed_addr constant [1 x i64] [i64 2]
// CHECK-DAG: [[MAPT2:@.+]] = private unnamed_addr constant [1 x i64] [i64 800] // CHECK-DAG: [[MAPT2:@.+]] = private unnamed_addr constant [1 x i64] [i64 800]
// CHECK-DAG: [[SIZET3:@.+]] = private unnamed_addr constant [2 x i64] [i64 4, i64 2] // CHECK-DAG: [[SIZET3:@.+]] = private unnamed_addr constant [2 x i64] [i64 4, i64 2]
// CHECK-DAG: [[MAPT3:@.+]] = private unnamed_addr constant [2 x i64] [i64 800, i64 800] // CHECK-DAG: [[MAPT3:@.+]] = private unnamed_addr constant [2 x i64] [i64 800, i64 800]
// CHECK-DAG: [[MAPT4:@.+]] = private unnamed_addr constant [10 x i64] [i64 800, i64 547, i64 800, i64 547, i64 547, i64 800, i64 800, i64 547, i64 547, i64 800] // CHECK-DAG: [[MAPT4:@.+]] = private unnamed_addr constant [10 x i64] [i64 800, i64 547, i64 800, i64 547, i64 547, i64 800, i64 800, i64 547, i64 547, i64 800]
// CHECK-DAG: [[MAPT5:@.+]] = private unnamed_addr constant [6 x i64] [i64 32, i64 281474976711171, i64 800, i64 800, i64 800, i64 547] // CHECK-DAG: [[MAPT5:@.+]] = private unnamed_addr constant [5 x i64] [i64 547, i64 800, i64 800, i64 800, i64 547]
// CHECK-DAG: [[SIZET6:@.+]] = private unnamed_addr constant [5 x i64] [i64 4, i64 4, i64 2, i64 1, i64 40] // CHECK-DAG: [[SIZET6:@.+]] = private unnamed_addr constant [5 x i64] [i64 4, i64 4, i64 2, i64 1, i64 40]
// CHECK-DAG: [[MAPT6:@.+]] = private unnamed_addr constant [5 x i64] [i64 800, i64 800, i64 800, i64 800, i64 547] // CHECK-DAG: [[MAPT6:@.+]] = private unnamed_addr constant [5 x i64] [i64 800, i64 800, i64 800, i64 800, i64 547]
// CHECK-DAG: [[SIZET7:@.+]] = private unnamed_addr constant [3 x i64] [i64 4, i64 2, i64 40] // CHECK-DAG: [[SIZET7:@.+]] = private unnamed_addr constant [3 x i64] [i64 4, i64 2, i64 40]
// CHECK-DAG: [[MAPT7:@.+]] = private unnamed_addr constant [3 x i64] [i64 800, i64 800, i64 547] // CHECK-DAG: [[MAPT7:@.+]] = private unnamed_addr constant [3 x i64] [i64 800, i64 800, i64 547]
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
▲ Show 20 Lines • Show All 502 Lines • ▼ Show 20 Lines
// CHECK: br i1 [[IF]], label %[[TRY:[^,]+]], label %[[FAIL:[^,]+]] // CHECK: br i1 [[IF]], label %[[TRY:[^,]+]], label %[[FAIL:[^,]+]]
// CHECK: [[TRY]] // CHECK: [[TRY]]
// We capture 2 VLA sizes in this target region // We capture 2 VLA sizes in this target region
// CHECK: [[CELEMSIZE2:%.+]] = mul nuw i[[SZ]] 2, [[VLA0:%.+]] // CHECK: [[CELEMSIZE2:%.+]] = mul nuw i[[SZ]] 2, [[VLA0:%.+]]
// CHECK-64: [[CSIZE:%.+]] = mul nuw i[[SZ]] [[CELEMSIZE2]], 2 // CHECK-64: [[CSIZE:%.+]] = mul nuw i[[SZ]] [[CELEMSIZE2]], 2
// CHECK-32: [[CSSZSIZE:%.+]] = mul nuw i[[SZ]] [[CELEMSIZE2]], 2 // CHECK-32: [[CSSZSIZE:%.+]] = mul nuw i[[SZ]] [[CELEMSIZE2]], 2
// CHECK-32: [[CSIZE:%.+]] = sext i32 [[CSSZSIZE]] to i64 // CHECK-32: [[CSIZE:%.+]] = sext i32 [[CSSZSIZE]] to i64
// CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 6, i8** [[BPR:%[^,]+]], i8** [[PR:%[^,]+]], i64* [[SR:%[^,]+]], i64* getelementptr inbounds ([6 x i64], [6 x i64]* [[MAPT5]], i32 0, i32 0), i8** null, i32 0, i32 0) // CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 5, i8** [[BPR:%[^,]+]], i8** [[PR:%[^,]+]], i64* [[SR:%[^,]+]], i64* getelementptr inbounds ([5 x i64], [5 x i64]* [[MAPT5]], i32 0, i32 0), i8** null, i32 0, i32 0)
// CHECK-DAG: [[BPR]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP:%.+]], i32 0, i32 0 // CHECK-DAG: [[BPR]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP:%.+]], i32 0, i32 0
// CHECK-DAG: [[PR]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P:%.+]], i32 0, i32 0 // CHECK-DAG: [[PR]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P:%.+]], i32 0, i32 0
// CHECK-DAG: [[SR]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S:%.+]], i32 0, i32 0 // CHECK-DAG: [[SR]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S:%.+]], i32 0, i32 0
// CHECK-DAG: [[SADDR0:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX0:[0-9]+]] // CHECK-DAG: [[SADDR0:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX0:[0-9]+]]
// CHECK-DAG: [[BPADDR0:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX0]] // CHECK-DAG: [[BPADDR0:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX0]]
// CHECK-DAG: [[PADDR0:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX0]] // CHECK-DAG: [[PADDR0:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX0]]
// CHECK-DAG: [[SADDR1:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX1:[0-9]+]] // CHECK-DAG: [[SADDR1:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX1:[0-9]+]]
// CHECK-DAG: [[BPADDR1:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX1]] // CHECK-DAG: [[BPADDR1:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX1]]
// CHECK-DAG: [[PADDR1:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX1]] // CHECK-DAG: [[PADDR1:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX1]]
// CHECK-DAG: [[SADDR2:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX2:[0-9]+]] // CHECK-DAG: [[SADDR2:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX2:[0-9]+]]
// CHECK-DAG: [[BPADDR2:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX2]] // CHECK-DAG: [[BPADDR2:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX2]]
// CHECK-DAG: [[PADDR2:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX2]] // CHECK-DAG: [[PADDR2:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX2]]
// CHECK-DAG: [[SADDR3:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX3:[0-9]+]] // CHECK-DAG: [[SADDR3:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX3:[0-9]+]]
// CHECK-DAG: [[BPADDR3:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX3]] // CHECK-DAG: [[BPADDR3:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX3]]
// CHECK-DAG: [[PADDR3:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX3]] // CHECK-DAG: [[PADDR3:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX3]]
// CHECK-DAG: [[SADDR4:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX3:[0-9]+]] // CHECK-DAG: [[SADDR4:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX3:[0-9]+]]
// CHECK-DAG: [[BPADDR4:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX3]] // CHECK-DAG: [[BPADDR4:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX3]]
// CHECK-DAG: [[PADDR4:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX3]] // CHECK-DAG: [[PADDR4:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX3]]
// CHECK-DAG: [[SADDR5:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX3:[0-9]+]]
// CHECK-DAG: [[BPADDR5:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX3]]
// CHECK-DAG: [[PADDR5:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX3]]
// The names below are not necessarily consistent with the names used for the // The names below are not necessarily consistent with the names used for the
// addresses above as some are repeated. // addresses above as some are repeated.
// CHECK-DAG: store i[[SZ]] [[VLA0]], i[[SZ]]* [[CBPADDR0:%.+]], // CHECK-DAG: store i[[SZ]] [[VLA0]], i[[SZ]]* [[CBPADDR0:%.+]],
// CHECK-DAG: store i[[SZ]] [[VLA0]], i[[SZ]]* [[CPADDR0:%.+]], // CHECK-DAG: store i[[SZ]] [[VLA0]], i[[SZ]]* [[CPADDR0:%.+]],
// CHECK-DAG: [[CBPADDR0]] = bitcast i8** {{%[^,]+}} to i[[SZ]]* // CHECK-DAG: [[CBPADDR0]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// CHECK-DAG: [[CPADDR0]] = bitcast i8** {{%[^,]+}} to i[[SZ]]* // CHECK-DAG: [[CPADDR0]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// CHECK-DAG: store i64 {{4|8}}, i64* {{%[^,]+}} // CHECK-DAG: store i64 {{4|8}}, i64* {{%[^,]+}}
Show All 11 Lines
// CHECK-DAG: store i64 4, i64* {{%[^,]+}} // CHECK-DAG: store i64 4, i64* {{%[^,]+}}
// CHECK-DAG: store [[S1]]* %{{.+}}, [[S1]]** [[CBPADDR3:%.+]], // CHECK-DAG: store [[S1]]* %{{.+}}, [[S1]]** [[CBPADDR3:%.+]],
// CHECK-DAG: store double* %{{.+}}, double** [[CPADDR3:%.+]], // CHECK-DAG: store double* %{{.+}}, double** [[CPADDR3:%.+]],
// CHECK-DAG: [[CBPADDR3]] = bitcast i8** {{%[^,]+}} to [[S1]]** // CHECK-DAG: [[CBPADDR3]] = bitcast i8** {{%[^,]+}} to [[S1]]**
// CHECK-DAG: [[CPADDR3]] = bitcast i8** {{%[^,]+}} to double** // CHECK-DAG: [[CPADDR3]] = bitcast i8** {{%[^,]+}} to double**
// CHECK-DAG: store i64 8, i64* {{%[^,]+}} // CHECK-DAG: store i64 8, i64* {{%[^,]+}}
// CHECK-DAG: store [[S1]]* %{{.+}}, [[S1]]** [[CBPADDR4:%.+]], // CHECK-DAG: store i16* %{{.+}}, i16** [[CBPADDR4:%.+]],
// CHECK-DAG: store double* %{{.+}}, double** [[CPADDR4:%.+]], // CHECK-DAG: store i16* %{{.+}}, i16** [[CPADDR4:%.+]],
// CHECK-DAG: [[CBPADDR4]] = bitcast i8** {{%[^,]+}} to [[S1]]** // CHECK-DAG: [[CBPADDR4]] = bitcast i8** {{%[^,]+}} to i16**
// CHECK-DAG: [[CPADDR4]] = bitcast i8** {{%[^,]+}} to double** // CHECK-DAG: [[CPADDR4]] = bitcast i8** {{%[^,]+}} to i16**
// CHECK-DAG: store i64 8, i64* {{%[^,]+}}
// CHECK-DAG: store i16* %{{.+}}, i16** [[CBPADDR5:%.+]],
// CHECK-DAG: store i16* %{{.+}}, i16** [[CPADDR5:%.+]],
// CHECK-DAG: [[CBPADDR5]] = bitcast i8** {{%[^,]+}} to i16**
// CHECK-DAG: [[CPADDR5]] = bitcast i8** {{%[^,]+}} to i16**
// CHECK-DAG: store i64 [[CSIZE]], i64* {{%[^,]+}} // CHECK-DAG: store i64 [[CSIZE]], i64* {{%[^,]+}}
// CHECK-NEXT: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0 // CHECK-NEXT: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0
// CHECK-NEXT: br i1 [[ERROR]], label %[[FAIL:[^,]+]], label %[[END:[^,]+]] // CHECK-NEXT: br i1 [[ERROR]], label %[[FAIL:[^,]+]], label %[[END:[^,]+]]
// CHECK: [[FAIL]] // CHECK: [[FAIL]]
// CHECK: call void [[HVT7:@.+]]({{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}) // CHECK: call void [[HVT7:@.+]]({{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}})
// CHECK-NEXT: br label %[[END]] // CHECK-NEXT: br label %[[END]]
▲ Show 20 Lines • Show All 219 Lines • Show Last 20 Lines

clang/test/OpenMP/target_teams_distribute_simd_codegen.cpp

Show First 20 Lines • Show All 85 Lines • ▼ Show 20 Lines
// CHECK-DAG: [[SIZET:@.+]] = private unnamed_addr constant [3 x i64] [i64 2, i64 4, i64 4] // CHECK-DAG: [[SIZET:@.+]] = private unnamed_addr constant [3 x i64] [i64 2, i64 4, i64 4]
// CHECK-DAG: [[MAPT:@.+]] = private unnamed_addr constant [3 x i64] [i64 288, i64 800, i64 800] // CHECK-DAG: [[MAPT:@.+]] = private unnamed_addr constant [3 x i64] [i64 288, i64 800, i64 800]
// CHECK-DAG: [[SIZET2:@.+]] = private unnamed_addr constant [1 x i64] [i64 2] // CHECK-DAG: [[SIZET2:@.+]] = private unnamed_addr constant [1 x i64] [i64 2]
// CHECK-DAG: [[MAPT2:@.+]] = private unnamed_addr constant [1 x i64] [i64 800] // CHECK-DAG: [[MAPT2:@.+]] = private unnamed_addr constant [1 x i64] [i64 800]
// CHECK-DAG: [[SIZET3:@.+]] = private unnamed_addr constant [2 x i64] [i64 4, i64 2] // CHECK-DAG: [[SIZET3:@.+]] = private unnamed_addr constant [2 x i64] [i64 4, i64 2]
// CHECK-DAG: [[MAPT3:@.+]] = private unnamed_addr constant [2 x i64] [i64 800, i64 800] // CHECK-DAG: [[MAPT3:@.+]] = private unnamed_addr constant [2 x i64] [i64 800, i64 800]
// CHECK-DAG: [[MAPT4:@.+]] = private unnamed_addr constant [9 x i64] [i64 800, i64 547, i64 800, i64 547, i64 547, i64 800, i64 800, i64 547, i64 547] // CHECK-DAG: [[MAPT4:@.+]] = private unnamed_addr constant [9 x i64] [i64 800, i64 547, i64 800, i64 547, i64 547, i64 800, i64 800, i64 547, i64 547]
// OMP45-DAG: [[MAPT5:@.+]] = private unnamed_addr constant [6 x i64] [i64 32, i64 281474976711171, i64 800, i64 800, i64 800, i64 547] // OMP45-DAG: [[MAPT5:@.+]] = private unnamed_addr constant [5 x i64] [i64 547, i64 800, i64 800, i64 800, i64 547]
// OMP50-DAG: [[MAPT5:@.+]] = private unnamed_addr constant [7 x i64] [i64 32, i64 281474976711171, i64 800, i64 800, i64 800, i64 547, i64 800] // OMP50-DAG: [[MAPT5:@.+]] = private unnamed_addr constant [6 x i64] [i64 547, i64 800, i64 800, i64 800, i64 547, i64 800]
// CHECK-DAG: [[SIZET6:@.+]] = private unnamed_addr constant [5 x i64] [i64 4, i64 4, i64 2, i64 1, i64 40] // CHECK-DAG: [[SIZET6:@.+]] = private unnamed_addr constant [5 x i64] [i64 4, i64 4, i64 2, i64 1, i64 40]
// CHECK-DAG: [[MAPT6:@.+]] = private unnamed_addr constant [5 x i64] [i64 800, i64 800, i64 800, i64 800, i64 547] // CHECK-DAG: [[MAPT6:@.+]] = private unnamed_addr constant [5 x i64] [i64 800, i64 800, i64 800, i64 800, i64 547]
// CHECK-DAG: [[SIZET7:@.+]] = private unnamed_addr constant [3 x i64] [i64 4, i64 2, i64 40] // CHECK-DAG: [[SIZET7:@.+]] = private unnamed_addr constant [3 x i64] [i64 4, i64 2, i64 40]
// CHECK-DAG: [[MAPT7:@.+]] = private unnamed_addr constant [3 x i64] [i64 800, i64 800, i64 547] // CHECK-DAG: [[MAPT7:@.+]] = private unnamed_addr constant [3 x i64] [i64 800, i64 800, i64 547]
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0 // CHECK-DAG: @{{.*}} = weak constant i8 0
▲ Show 20 Lines • Show All 499 Lines • ▼ Show 20 Lines
// CHECK: br i1 [[IF]], label %[[TRY:[^,]+]], label %[[FAIL:[^,]+]] // CHECK: br i1 [[IF]], label %[[TRY:[^,]+]], label %[[FAIL:[^,]+]]
// CHECK: [[TRY]] // CHECK: [[TRY]]
// We capture 2 VLA sizes in this target region // We capture 2 VLA sizes in this target region
// CHECK: [[CELEMSIZE2:%.+]] = mul nuw i[[SZ]] 2, [[VLA0:%.+]] // CHECK: [[CELEMSIZE2:%.+]] = mul nuw i[[SZ]] 2, [[VLA0:%.+]]
// CHECK-64: [[CSIZE:%.+]] = mul nuw i64 [[CELEMSIZE2]], 2 // CHECK-64: [[CSIZE:%.+]] = mul nuw i64 [[CELEMSIZE2]], 2
// CHECK-32: [[CSZSIZE:%.+]] = mul nuw i32 [[CELEMSIZE2]], 2 // CHECK-32: [[CSZSIZE:%.+]] = mul nuw i32 [[CELEMSIZE2]], 2
// CHECK-32: [[CSIZE:%.+]] = sext i32 [[CSZSIZE]] to i64 // CHECK-32: [[CSIZE:%.+]] = sext i32 [[CSZSIZE]] to i64
// OMP45-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 6, i8** [[BPR:%[^,]+]], i8** [[PR:%[^,]+]], i64* [[SR:%[^,]+]], i64* getelementptr inbounds ([6 x i64], [6 x i64]* [[MAPT5]], i32 0, i32 0), i8** null, i32 0, i32 1) // OMP45-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 5, i8** [[BPR:%[^,]+]], i8** [[PR:%[^,]+]], i64* [[SR:%[^,]+]], i64* getelementptr inbounds ([5 x i64], [5 x i64]* [[MAPT5]], i32 0, i32 0), i8** null, i32 0, i32 1)
// OMP45-DAG: [[BPR]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP:%.+]], i32 0, i32 0 // OMP45-DAG: [[BPR]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP:%.+]], i32 0, i32 0
// OMP45-DAG: [[PR]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P:%.+]], i32 0, i32 0 // OMP45-DAG: [[PR]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P:%.+]], i32 0, i32 0
// OMP45-DAG: [[SR]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S:%.+]], i32 0, i32 0 // OMP45-DAG: [[SR]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S:%.+]], i32 0, i32 0
// OMP45-DAG: [[SADDR0:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX0:[0-9]+]] // OMP45-DAG: [[SADDR0:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX0:[0-9]+]]
// OMP45-DAG: [[BPADDR0:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX0]] // OMP45-DAG: [[BPADDR0:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX0]]
// OMP45-DAG: [[PADDR0:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX0]] // OMP45-DAG: [[PADDR0:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX0]]
// OMP45-DAG: [[SADDR1:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX1:[0-9]+]] // OMP45-DAG: [[SADDR1:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX1:[0-9]+]]
// OMP45-DAG: [[BPADDR1:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX1]] // OMP45-DAG: [[BPADDR1:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX1]]
// OMP45-DAG: [[PADDR1:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX1]] // OMP45-DAG: [[PADDR1:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX1]]
// OMP45-DAG: [[SADDR2:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX2:[0-9]+]] // OMP45-DAG: [[SADDR2:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX2:[0-9]+]]
// OMP45-DAG: [[BPADDR2:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX2]] // OMP45-DAG: [[BPADDR2:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX2]]
// OMP45-DAG: [[PADDR2:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX2]] // OMP45-DAG: [[PADDR2:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX2]]
// OMP45-DAG: [[SADDR3:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX3:[0-9]+]] // OMP45-DAG: [[SADDR3:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX3:[0-9]+]]
// OMP45-DAG: [[BPADDR3:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX3]] // OMP45-DAG: [[BPADDR3:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX3]]
// OMP45-DAG: [[PADDR3:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX3]] // OMP45-DAG: [[PADDR3:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX3]]
// OMP45-DAG: [[SADDR4:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX4:[0-9]+]] // OMP45-DAG: [[SADDR4:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX4:[0-9]+]]
// OMP45-DAG: [[BPADDR4:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX4]] // OMP45-DAG: [[BPADDR4:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX4]]
// OMP45-DAG: [[PADDR4:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX4]] // OMP45-DAG: [[PADDR4:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX4]]
// OMP45-DAG: [[SADDR5:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX5:[0-9]+]] // OMP50-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 6, i8** [[BPR:%[^,]+]], i8** [[PR:%[^,]+]], i64* [[SR:%[^,]+]], i64* getelementptr inbounds ([6 x i64], [6 x i64]* [[MAPT5]], i32 0, i32 0), i8** null, i32 0, i32 1)
// OMP45-DAG: [[BPADDR5:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX5]] // OMP50-DAG: [[BPR]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP:%.+]], i32 0, i32 0
// OMP45-DAG: [[PADDR5:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX5]] // OMP50-DAG: [[PR]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P:%.+]], i32 0, i32 0
// OMP50-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 7, i8** [[BPR:%[^,]+]], i8** [[PR:%[^,]+]], i64* [[SR:%[^,]+]], i64* getelementptr inbounds ([7 x i64], [7 x i64]* [[MAPT5]], i32 0, i32 0), i8** null, i32 0, i32 1) // OMP50-DAG: [[SR]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S:%.+]], i32 0, i32 0
// OMP50-DAG: [[BPR]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[BP:%.+]], i32 0, i32 0 // OMP50-DAG: [[SADDR0:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX0:[0-9]+]]
// OMP50-DAG: [[PR]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[P:%.+]], i32 0, i32 0 // OMP50-DAG: [[BPADDR0:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX0]]
// OMP50-DAG: [[SR]] = getelementptr inbounds [7 x i64], [7 x i64]* [[S:%.+]], i32 0, i32 0 // OMP50-DAG: [[PADDR0:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX0]]
// OMP50-DAG: [[SADDR0:%.+]] = getelementptr inbounds [7 x i64], [7 x i64]* [[S]], i32 [[IDX0:[0-9]+]] // OMP50-DAG: [[SADDR1:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX1:[0-9]+]]
// OMP50-DAG: [[BPADDR0:%.+]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[BP]], i32 [[IDX0]] // OMP50-DAG: [[BPADDR1:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX1]]
// OMP50-DAG: [[PADDR0:%.+]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[P]], i32 [[IDX0]] // OMP50-DAG: [[PADDR1:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX1]]
// OMP50-DAG: [[SADDR1:%.+]] = getelementptr inbounds [7 x i64], [7 x i64]* [[S]], i32 [[IDX1:[0-9]+]] // OMP50-DAG: [[SADDR2:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX2:[0-9]+]]
// OMP50-DAG: [[BPADDR1:%.+]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[BP]], i32 [[IDX1]] // OMP50-DAG: [[BPADDR2:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX2]]
// OMP50-DAG: [[PADDR1:%.+]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[P]], i32 [[IDX1]] // OMP50-DAG: [[PADDR2:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX2]]
// OMP50-DAG: [[SADDR2:%.+]] = getelementptr inbounds [7 x i64], [7 x i64]* [[S]], i32 [[IDX2:[0-9]+]] // OMP50-DAG: [[SADDR3:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX3:[0-9]+]]
// OMP50-DAG: [[BPADDR2:%.+]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[BP]], i32 [[IDX2]] // OMP50-DAG: [[BPADDR3:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX3]]
// OMP50-DAG: [[PADDR2:%.+]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[P]], i32 [[IDX2]] // OMP50-DAG: [[PADDR3:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX3]]
// OMP50-DAG: [[SADDR3:%.+]] = getelementptr inbounds [7 x i64], [7 x i64]* [[S]], i32 [[IDX3:[0-9]+]] // OMP50-DAG: [[SADDR4:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX4:[0-9]+]]
// OMP50-DAG: [[BPADDR3:%.+]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[BP]], i32 [[IDX3]] // OMP50-DAG: [[BPADDR4:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX4]]
// OMP50-DAG: [[PADDR3:%.+]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[P]], i32 [[IDX3]] // OMP50-DAG: [[PADDR4:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX4]]
// OMP50-DAG: [[SADDR4:%.+]] = getelementptr inbounds [7 x i64], [7 x i64]* [[S]], i32 [[IDX4:[0-9]+]] // OMP50-DAG: [[SADDR5:%.+]] = getelementptr inbounds [6 x i64], [6 x i64]* [[S]], i32 [[IDX5:[0-9]+]]
// OMP50-DAG: [[BPADDR4:%.+]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[BP]], i32 [[IDX4]] // OMP50-DAG: [[BPADDR5:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[BP]], i32 [[IDX5]]
// OMP50-DAG: [[PADDR4:%.+]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[P]], i32 [[IDX4]] // OMP50-DAG: [[PADDR5:%.+]] = getelementptr inbounds [6 x i8*], [6 x i8*]* [[P]], i32 [[IDX5]]
// OMP50-DAG: [[SADDR5:%.+]] = getelementptr inbounds [7 x i64], [7 x i64]* [[S]], i32 [[IDX5:[0-9]+]]
// OMP50-DAG: [[BPADDR5:%.+]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[BP]], i32 [[IDX5]]
// OMP50-DAG: [[PADDR5:%.+]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[P]], i32 [[IDX5]]
// OMP50-DAG: [[SADDR6:%.+]] = getelementptr inbounds [7 x i64], [7 x i64]* [[S]], i32 [[IDX6:[0-9]+]]
// OMP50-DAG: [[BPADDR6:%.+]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[BP]], i32 [[IDX6]]
// OMP50-DAG: [[PADDR6:%.+]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[P]], i32 [[IDX6]]
// The names below are not necessarily consistent with the names used for the // The names below are not necessarily consistent with the names used for the
// addresses above as some are repeated. // addresses above as some are repeated.
// CHECK-DAG: store [[S1]]* %{{.+}}, [[S1]]** [[CBPADDR0:%.+]], // CHECK-DAG: store [[S1]]* %{{.+}}, [[S1]]** [[CBPADDR0:%.+]],
// CHECK-DAG: store double* %{{.+}}, double** [[CPADDR0:%.+]], // CHECK-DAG: store double* %{{.+}}, double** [[CPADDR0:%.+]],
// CHECK-DAG: [[CBPADDR0]] = bitcast i8** {{%[^,]+}} to [[S1]]** // CHECK-DAG: [[CBPADDR0]] = bitcast i8** {{%[^,]+}} to [[S1]]**
// CHECK-DAG: [[CPADDR0]] = bitcast i8** {{%[^,]+}} to double** // CHECK-DAG: [[CPADDR0]] = bitcast i8** {{%[^,]+}} to double**
// CHECK-DAG: store i64 %{{.+}}, i64* {{%[^,]+}}
// CHECK-DAG: store [[S1]]* %{{.+}}, [[S1]]** [[CBPADDR1:%.+]],
// CHECK-DAG: store double* %{{.+}}, double** [[CPADDR1:%.+]],
// CHECK-DAG: [[CBPADDR1]] = bitcast i8** {{%[^,]+}} to [[S1]]**
// CHECK-DAG: [[CPADDR1]] = bitcast i8** {{%[^,]+}} to double**
// CHECK-DAG: store i64 {{8|4}}, i64* {{%[^,]+}} // CHECK-DAG: store i64 {{8|4}}, i64* {{%[^,]+}}
// CHECK-DAG: store i[[SZ]] [[B_CVAL]], i[[SZ]]* [[CBPADDR2:%.+]], // CHECK-DAG: store i[[SZ]] [[B_CVAL]], i[[SZ]]* [[CBPADDR1:%.+]],
// CHECK-DAG: store i[[SZ]] [[B_CVAL]], i[[SZ]]* [[CPADDR2:%.+]], // CHECK-DAG: store i[[SZ]] [[B_CVAL]], i[[SZ]]* [[CPADDR1:%.+]],
// CHECK-DAG: [[CBPADDR1]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// CHECK-DAG: [[CPADDR1]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// CHECK-DAG: store i64 4, i64* {{%[^,]+}}
// CHECK-DAG: store i[[SZ]] 2, i[[SZ]]* [[CBPADDR2:%.+]],
// CHECK-DAG: store i[[SZ]] 2, i[[SZ]]* [[CPADDR2:%.+]],
// CHECK-DAG: [[CBPADDR2]] = bitcast i8** {{%[^,]+}} to i[[SZ]]* // CHECK-DAG: [[CBPADDR2]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// CHECK-DAG: [[CPADDR2]] = bitcast i8** {{%[^,]+}} to i[[SZ]]* // CHECK-DAG: [[CPADDR2]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// CHECK-DAG: store i64 4, i64* {{%[^,]+}} // CHECK-DAG: store i64 {{4|8}}, i64* {{%[^,]+}}
// CHECK-DAG: store i[[SZ]] 2, i[[SZ]]* [[CBPADDR3:%.+]], // CHECK-DAG: store i[[SZ]] [[VLA0]], i[[SZ]]* [[CBPADDR3:%.+]],
// CHECK-DAG: store i[[SZ]] 2, i[[SZ]]* [[CPADDR3:%.+]], // CHECK-DAG: store i[[SZ]] [[VLA0]], i[[SZ]]* [[CPADDR3:%.+]],
// CHECK-DAG: [[CBPADDR3]] = bitcast i8** {{%[^,]+}} to i[[SZ]]* // CHECK-DAG: [[CBPADDR3]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// CHECK-DAG: [[CPADDR3]] = bitcast i8** {{%[^,]+}} to i[[SZ]]* // CHECK-DAG: [[CPADDR3]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// CHECK-DAG: store i64 {{4|8}}, i64* {{%[^,]+}} // CHECK-DAG: store i64 {{4|8}}, i64* {{%[^,]+}}
// CHECK-DAG: store i[[SZ]] [[VLA0]], i[[SZ]]* [[CBPADDR4:%.+]], // CHECK-DAG: store i16* %{{.+}}, i16** [[CBPADDR4:%.+]],
// CHECK-DAG: store i[[SZ]] [[VLA0]], i[[SZ]]* [[CPADDR4:%.+]], // CHECK-DAG: store i16* %{{.+}}, i16** [[CPADDR4:%.+]],
// CHECK-DAG: [[CBPADDR4]] = bitcast i8** {{%[^,]+}} to i[[SZ]]* // CHECK-DAG: [[CBPADDR4]] = bitcast i8** {{%[^,]+}} to i16**
// CHECK-DAG: [[CPADDR4]] = bitcast i8** {{%[^,]+}} to i[[SZ]]* // CHECK-DAG: [[CPADDR4]] = bitcast i8** {{%[^,]+}} to i16**
// CHECK-DAG: store i64 {{4|8}}, i64* {{%[^,]+}}
// CHECK-DAG: store i16* %{{.+}}, i16** [[CBPADDR5:%.+]],
// CHECK-DAG: store i16* %{{.+}}, i16** [[CPADDR5:%.+]],
// CHECK-DAG: [[CBPADDR5]] = bitcast i8** {{%[^,]+}} to i16**
// CHECK-DAG: [[CPADDR5]] = bitcast i8** {{%[^,]+}} to i16**
// CHECK-DAG: store i64 [[CSIZE]], i64* {{%[^,]+}} // CHECK-DAG: store i64 [[CSIZE]], i64* {{%[^,]+}}
// OMP50-DAG: store i[[SZ]] [[IF_C:%.+]], i[[SZ]]* [[CBPADDR6:%.+]], // OMP50-DAG: store i[[SZ]] [[IF_C:%.+]], i[[SZ]]* [[CBPADDR5:%.+]],
// OMP50-DAG: store i[[SZ]] [[IF_C]], i[[SZ]]* [[CPADDR6:%.+]], // OMP50-DAG: store i[[SZ]] [[IF_C]], i[[SZ]]* [[CPADDR5:%.+]],
// OMP50-DAG: [[CBPADDR6]] = bitcast i8** {{%[^,]+}} to i[[SZ]]* // OMP50-DAG: [[CBPADDR5]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// OMP50-DAG: [[CPADDR6]] = bitcast i8** {{%[^,]+}} to i[[SZ]]* // OMP50-DAG: [[CPADDR5]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// OMP50-DAG: store i64 1, i64* {{%[^,]+}} // OMP50-DAG: store i64 1, i64* {{%[^,]+}}
// CHECK-NEXT: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0 // CHECK-NEXT: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0
// CHECK-NEXT: br i1 [[ERROR]], label %[[FAIL:[^,]+]], label %[[END:[^,]+]] // CHECK-NEXT: br i1 [[ERROR]], label %[[FAIL:[^,]+]], label %[[END:[^,]+]]
// CHECK: [[FAIL]] // CHECK: [[FAIL]]
// OMP45: call void [[HVT7:@.+]]({{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}) // OMP45: call void [[HVT7:@.+]]({{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}})
// OMP50: call void [[HVT7:@.+]]({{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}) // OMP50: call void [[HVT7:@.+]]({{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}})
▲ Show 20 Lines • Show All 232 Lines • Show Last 20 Lines

clang/test/OpenMP/target_teams_num_teams_codegen.cpp

Show First 20 Lines • Show All 183 Lines • ▼ Show 20 Lines
// CHECK: store i32 [[SUB]], i32* [[CAPE_ADDR:%.+]], align // CHECK: store i32 [[SUB]], i32* [[CAPE_ADDR:%.+]], align
// CHECK: [[CEV:%.+]] = load i32, i32* [[CAPE_ADDR]], align // CHECK: [[CEV:%.+]] = load i32, i32* [[CAPE_ADDR]], align
// CHECK-64: [[CONV:%.+]] = bitcast i[[SZ]]* [[CAPEC_ADDR:%.+]] to i32* // CHECK-64: [[CONV:%.+]] = bitcast i[[SZ]]* [[CAPEC_ADDR:%.+]] to i32*
// CHECK-64: store i32 [[CEV]], i32* [[CONV]], align // CHECK-64: store i32 [[CEV]], i32* [[CONV]], align
// CHECK-32: store i32 [[CEV]], i32* [[CAPEC_ADDR:%.+]], align // CHECK-32: store i32 [[CEV]], i32* [[CAPEC_ADDR:%.+]], align
// CHECK: [[ARG:%.+]] = load i[[SZ]], i[[SZ]]* [[CAPEC_ADDR]], align // CHECK: [[ARG:%.+]] = load i[[SZ]], i[[SZ]]* [[CAPEC_ADDR]], align
// CHECK: [[TEAMS:%.+]] = load i32, i32* [[CAPE_ADDR]], align // CHECK: [[TEAMS:%.+]] = load i32, i32* [[CAPE_ADDR]], align
// //
// CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 4, {{.*}}, i8** null, i32 [[TEAMS]], i32 0) // CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 3, {{.*}}, i8** null, i32 [[TEAMS]], i32 0)
// CHECK: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0 // CHECK: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0
// CHECK: br i1 [[ERROR]], label %[[FAIL:.+]], label %[[END:[^,]+]] // CHECK: br i1 [[ERROR]], label %[[FAIL:.+]], label %[[END:[^,]+]]
// //
// CHECK: [[FAIL]] // CHECK: [[FAIL]]
// CHECK: call void [[HVT1:@.+]]([[S1]]* {{%.+}}, i[[SZ]] {{%.+}}, i[[SZ]] [[ARG]]) // CHECK: call void [[HVT1:@.+]]([[S1]]* {{%.+}}, i[[SZ]] {{%.+}}, i[[SZ]] [[ARG]])
// CHECK: br label {{%?}}[[END]] // CHECK: br label {{%?}}[[END]]
// CHECK: [[END]] // CHECK: [[END]]
// //
// //
// //
// CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 2, {{.+}}, i8** null, i32 1024, i32 0) // CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 1, {{.+}}, i8** null, i32 1024, i32 0)
// CHECK: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0 // CHECK: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0
// CHECK: br i1 [[ERROR]], label %[[FAIL:.+]], label %[[END:[^,]+]] // CHECK: br i1 [[ERROR]], label %[[FAIL:.+]], label %[[END:[^,]+]]
// //
// CHECK: [[FAIL]] // CHECK: [[FAIL]]
// CHECK: call void [[HVT2:@.+]]([[S1]]* {{[^,]+}}) // CHECK: call void [[HVT2:@.+]]([[S1]]* {{[^,]+}})
// CHECK: br label {{%?}}[[END]] // CHECK: br label {{%?}}[[END]]
// CHECK: [[END]] // CHECK: [[END]]
// //
▲ Show 20 Lines • Show All 150 Lines • Show Last 20 Lines

clang/test/OpenMP/target_teams_thread_limit_codegen.cpp

Show First 20 Lines • Show All 183 Lines • ▼ Show 20 Lines
// CHECK: store i32 [[SUB]], i32* [[CAPE_ADDR:%.+]], align // CHECK: store i32 [[SUB]], i32* [[CAPE_ADDR:%.+]], align
// CHECK: [[CEV:%.+]] = load i32, i32* [[CAPE_ADDR]], align // CHECK: [[CEV:%.+]] = load i32, i32* [[CAPE_ADDR]], align
// CHECK-64: [[CONV:%.+]] = bitcast i[[SZ]]* [[CAPEC_ADDR:%.+]] to i32* // CHECK-64: [[CONV:%.+]] = bitcast i[[SZ]]* [[CAPEC_ADDR:%.+]] to i32*
// CHECK-64: store i32 [[CEV]], i32* [[CONV]], align // CHECK-64: store i32 [[CEV]], i32* [[CONV]], align
// CHECK-32: store i32 [[CEV]], i32* [[CAPEC_ADDR:%.+]], align // CHECK-32: store i32 [[CEV]], i32* [[CAPEC_ADDR:%.+]], align
// CHECK: [[ARG:%.+]] = load i[[SZ]], i[[SZ]]* [[CAPEC_ADDR]], align // CHECK: [[ARG:%.+]] = load i[[SZ]], i[[SZ]]* [[CAPEC_ADDR]], align
// CHECK: [[TL:%.+]] = load i32, i32* [[CAPE_ADDR]], align // CHECK: [[TL:%.+]] = load i32, i32* [[CAPE_ADDR]], align
// //
// CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 4, {{.*}}, i8** null, i32 0, i32 [[TL]]) // CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 3, {{.*}}, i8** null, i32 0, i32 [[TL]])
// CHECK: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0 // CHECK: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0
// CHECK: br i1 [[ERROR]], label %[[FAIL:.+]], label %[[END:[^,]+]] // CHECK: br i1 [[ERROR]], label %[[FAIL:.+]], label %[[END:[^,]+]]
// //
// CHECK: [[FAIL]] // CHECK: [[FAIL]]
// CHECK: call void [[HVT1:@.+]]([[S1]]* {{%.+}}, i[[SZ]] {{%.+}}, i[[SZ]] [[ARG]]) // CHECK: call void [[HVT1:@.+]]([[S1]]* {{%.+}}, i[[SZ]] {{%.+}}, i[[SZ]] [[ARG]])
// CHECK: br label {{%?}}[[END]] // CHECK: br label {{%?}}[[END]]
// CHECK: [[END]] // CHECK: [[END]]
// //
// //
// //
// CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 2, {{.+}}, i8** null, i32 0, i32 1024) // CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 1, {{.+}}, i8** null, i32 0, i32 1024)
// CHECK: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0 // CHECK: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0
// CHECK: br i1 [[ERROR]], label %[[FAIL:.+]], label %[[END:[^,]+]] // CHECK: br i1 [[ERROR]], label %[[FAIL:.+]], label %[[END:[^,]+]]
// //
// CHECK: [[FAIL]] // CHECK: [[FAIL]]
// CHECK: call void [[HVT2:@.+]]([[S1]]* {{[^,]+}}) // CHECK: call void [[HVT2:@.+]]([[S1]]* {{[^,]+}})
// CHECK: br label {{%?}}[[END]] // CHECK: br label {{%?}}[[END]]
// CHECK: [[END]] // CHECK: [[END]]
// //
▲ Show 20 Lines • Show All 163 Lines • Show Last 20 Lines

clang/test/OpenMP/teams_codegen.cpp

Show First 20 Lines • Show All 199 Lines • ▼ Show 20 Lines
template <typename T, int X, long long Y> template <typename T, int X, long long Y>
struct SS{ struct SS{
T a; T a;
float b; float b;
int foo(void) { int foo(void) {
int comp = 1; int comp = 1;
// CK3-DAG: call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 3, i8** %{{[^,]+}}, i8** %{{[^,]+}}, i{{64|32}}* %{{[^,]+}}, i64* {{.+}}@{{[^,]+}}, i32 0, i32 0), i8** null, i32 [[NT:%[^,]+]], i32 123) // CK3-DAG: call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 2, i8** %{{[^,]+}}, i8** %{{[^,]+}}, i64* {{.+}}@{{[^,]+}}, i32 0, i32 0), i64* {{.+}}@{{[^,]+}}, i32 0, i32 0), i8** null, i32 [[NT:%[^,]+]], i32 123)
// CK3-DAG: [[NT]] = load i32, i32* [[NTA:%[^,]+]], // CK3-DAG: [[NT]] = load i32, i32* [[NTA:%[^,]+]],
// CK3-DAG: [[NTA]] = getelementptr inbounds [[SSI]], [[SSI]]* [[NTB:%[^,]+]], i32 0, i32 0 // CK3-DAG: [[NTA]] = getelementptr inbounds [[SSI]], [[SSI]]* [[NTB:%[^,]+]], i32 0, i32 0
// CK3-DAG: [[NTB]] = load [[SSI]]*, [[SSI]]** %{{.+}}, // CK3-DAG: [[NTB]] = load [[SSI]]*, [[SSI]]** %{{.+}},
// CK3: call void @{{.+}}({{.+}} {{.+}}, {{.+}} {{.+}}) // CK3: call void @{{.+}}({{.+}} {{.+}}, {{.+}} {{.+}})
#pragma omp target #pragma omp target
#pragma omp teams num_teams(a) thread_limit(X) #pragma omp teams num_teams(a) thread_limit(X)
{ {
++comp; ++comp;
} }
// CK3-DAG: call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 3, i8** %{{[^,]+}}, i8** %{{[^,]+}}, i{{64|32}}* %{{[^,]+}}, i64* {{.+}}@{{[^,]+}}, i32 0, i32 0), i8** null, i32 456, i32 [[TL:%[^,]+]]) // CK3-DAG: call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 2, i8** %{{[^,]+}}, i8** %{{[^,]+}}, i64* {{.+}}@{{[^,]+}}, i32 0, i32 0), i64* {{.+}}@{{[^,]+}}, i32 0, i32 0), i8** null, i32 456, i32 [[TL:%[^,]+]])
// CK3-DAG: [[TL]] = add nsw i32 [[TLA:%[^,]+]], 123 // CK3-DAG: [[TL]] = add nsw i32 [[TLA:%[^,]+]], 123
// CK3-DAG: [[TLA]] = fptosi float [[TLB:%[^,]+]] to i32 // CK3-DAG: [[TLA]] = fptosi float [[TLB:%[^,]+]] to i32
// CK3-DAG: [[TLB]] = load float, float* [[TLC:%[^,]+]], // CK3-DAG: [[TLB]] = load float, float* [[TLC:%[^,]+]],
// CK3-DAG: [[TLC]] = getelementptr inbounds [[SSI]], [[SSI]]* [[THIS:%[^,]+]], i32 0, i32 1 // CK3-DAG: [[TLC]] = getelementptr inbounds [[SSI]], [[SSI]]* [[THIS:%[^,]+]], i32 0, i32 1
// CK3: call void @{{.+}}({{.+}} {{.+}}, {{.+}} {{.+}}) // CK3: call void @{{.+}}({{.+}} {{.+}}, {{.+}} {{.+}})
#pragma omp target #pragma omp target
▲ Show 20 Lines • Show All 195 Lines • Show Last 20 Lines

clang/test/OpenMP/teams_distribute_codegen.cpp

Show First 20 Lines • Show All 152 Lines • ▼ Show 20 Lines
template <typename T, int X, long long Y> template <typename T, int X, long long Y>
struct SS{ struct SS{
T a[X]; T a[X];
float b; float b;
// CK3: define {{.*}}i32 @{{.+}}foo{{.+}}( // CK3: define {{.*}}i32 @{{.+}}foo{{.+}}(
int foo(void) { int foo(void) {
// CK3: call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 2, i8** %{{[^,]+}}, i8** %{{[^,]+}}, i{{64|32}}* %{{.+}}, i64* {{.+}}@{{[^,]+}}, i32 0, i32 0), i8** null, i32 0, i32 0) // CK3: call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 1, i8** %{{[^,]+}}, i8** %{{[^,]+}}, i64* {{.+}}@{{[^,]+}}, i32 0, i32 0), i64* {{.+}}@{{[^,]+}}, i32 0, i32 0), i8** null, i32 0, i32 0)
// CK3: call void @[[OFFL1:.+]]([[SSI]]* %{{.+}}) // CK3: call void @[[OFFL1:.+]]([[SSI]]* %{{.+}})
#pragma omp target #pragma omp target
#pragma omp teams distribute #pragma omp teams distribute
for(int i = 0; i < X; i++) { for(int i = 0; i < X; i++) {
a[i] = (T)0; a[i] = (T)0;
} }
// outlined target region // outlined target region
▲ Show 20 Lines • Show All 105 Lines • Show Last 20 Lines

clang/test/OpenMP/teams_distribute_parallel_for_codegen.cpp

Show First 20 Lines • Show All 155 Lines • ▼ Show 20 Lines
template <typename T, int X, long long Y> template <typename T, int X, long long Y>
struct SS{ struct SS{
T a[X]; T a[X];
float b; float b;
// CK3: define {{.*}}i32 @{{.+}}foo{{.+}}( // CK3: define {{.*}}i32 @{{.+}}foo{{.+}}(
int foo(void) { int foo(void) {
// CK3: call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 2, i8** %{{[^,]+}}, i8** %{{[^,]+}}, i{{64|32}}* %{{.+}}, i64* {{.+}}@{{[^,]+}}, i32 0, i32 0), i8** null, i32 0, i32 0) // CK3: call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 1, i8** %{{[^,]+}}, i8** %{{[^,]+}}, i64* {{.+}}@{{[^,]+}}, i32 0, i32 0), i64* {{.+}}@{{[^,]+}}, i32 0, i32 0), i8** null, i32 0, i32 0)
// CK3: call void @[[OFFL1:.+]]([[SSI]]* %{{.+}}) // CK3: call void @[[OFFL1:.+]]([[SSI]]* %{{.+}})
#pragma omp target #pragma omp target
#pragma omp teams distribute parallel for #pragma omp teams distribute parallel for
for(int i = 0; i < X; i++) { for(int i = 0; i < X; i++) {
a[i] = (T)0; a[i] = (T)0;
} }
// outlined target region // outlined target region
▲ Show 20 Lines • Show All 108 Lines • Show Last 20 Lines

clang/test/OpenMP/teams_distribute_parallel_for_simd_codegen.cpp

Show First 20 Lines • Show All 163 Lines • ▼ Show 20 Lines
template <typename T, int X, long long Y> template <typename T, int X, long long Y>
struct SS{ struct SS{
T a[X]; T a[X];
float b; float b;
// CK3: define {{.*}}i32 @{{.+}}foo{{.+}}( // CK3: define {{.*}}i32 @{{.+}}foo{{.+}}(
int foo(void) { int foo(void) {
int i; int i;
// CK3: call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 3, i8** %{{[^,]+}}, i8** %{{[^,]+}}, i{{64|32}}* %{{[^,]+}}, i64* {{.+}}@{{[^,]+}}, i32 0, i32 0), i8** null // CK3: call i32 @__tgt_target_teams_mapper(i64 -1, i8* @{{[^,]+}}, i32 2, i8** %{{[^,]+}}, i8** %{{[^,]+}}, i64* {{.+}}@{{[^,]+}}, i32 0, i32 0), i64* {{.+}}@{{[^,]+}}, i32 0, i32 0), i8** null
// CK3: call void @[[OFFL1:.+]]([[SSI]]* %{{.+}}) // CK3: call void @[[OFFL1:.+]]([[SSI]]* %{{.+}})
#pragma omp target #pragma omp target
#pragma omp teams distribute parallel for simd safelen(4) aligned(a) linear(i) #pragma omp teams distribute parallel for simd safelen(4) aligned(a) linear(i)
for(i = 0; i < X; i++) { for(i = 0; i < X; i++) {
a[i] = (T)0; a[i] = (T)0;
} }
// outlined target region // outlined target region
▲ Show 20 Lines • Show All 116 Lines • Show Last 20 Lines