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

[OPENMP]Do not use OMP_MAP_TARGET_PARAM for data movement directives. 

OMP_MAP_TARGET_PARAM flag is used to mark the data that shoud be passed
as arguments to the target kernels, nothing else. But the compiler still
marks the data with OMP_MAP_TARGET_PARAM flags even if the data is
passed to the data movement directives, like target data, target update
etc. This flag is just ignored for this directives and the compiler does
not need to emit it.

Reviewed By: cchen

Differential Revision: https://reviews.llvm.org/D91261
This commit is contained in:
Alexey Bataev 2021-01-19 12:22:53 -08:00 committed by Alexey Bataev
parent cabe1b1124
commit b272698de7
14 changed files with 139 additions and 149 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

30
clang/lib/CodeGen/CGOpenMPRuntime.cpp
View File

@ -8233,7 +8233,7 @@ public:
MapFlagsArrayTy &CurTypes,
const StructRangeInfoTy &PartialStruct,
const ValueDecl *VD = nullptr,
bool NotTargetParams = false) const {
bool NotTargetParams = true) const {
if (CurTypes.size() == 1 &&
((CurTypes.back() & OMP_MAP_MEMBER_OF) != OMP_MAP_MEMBER_OF) &&
!PartialStruct.IsArraySection)
@ -8284,7 +8284,7 @@ public:
/// pair of the relevant declaration and index where it occurs is appended to
/// the device pointers info array.
void generateAllInfo(
MapCombinedInfoTy &CombinedInfo, bool NotTargetParams = false,
MapCombinedInfoTy &CombinedInfo,
const llvm::DenseSet<CanonicalDeclPtr<const Decl>> &SkipVarSet =
llvm::DenseSet<CanonicalDeclPtr<const Decl>>()) const {
// We have to process the component lists that relate with the same
@ -8420,9 +8420,7 @@ public:
UseDevicePtrCombinedInfo.Pointers.push_back(Ptr);
UseDevicePtrCombinedInfo.Sizes.push_back(
llvm::Constant::getNullValue(CGF.Int64Ty));
UseDevicePtrCombinedInfo.Types.push_back(
OMP_MAP_RETURN_PARAM |
(NotTargetParams ? OMP_MAP_NONE : OMP_MAP_TARGET_PARAM));
UseDevicePtrCombinedInfo.Types.push_back(OMP_MAP_RETURN_PARAM);
UseDevicePtrCombinedInfo.Mappers.push_back(nullptr);
}
}
@ -8490,19 +8488,13 @@ public:
CombinedInfo.Pointers.push_back(Ptr);
CombinedInfo.Sizes.push_back(
llvm::Constant::getNullValue(CGF.Int64Ty));
CombinedInfo.Types.push_back(
OMP_MAP_RETURN_PARAM |
(NotTargetParams ? OMP_MAP_NONE : OMP_MAP_TARGET_PARAM));
CombinedInfo.Types.push_back(OMP_MAP_RETURN_PARAM);
CombinedInfo.Mappers.push_back(nullptr);
}
}
}
for (const auto &M : Info) {
// We need to know when we generate information for the first component
// associated with a capture, because the mapping flags depend on it.
bool IsFirstComponentList = !NotTargetParams;
// Underlying variable declaration used in the map clause.
const ValueDecl *VD = std::get<0>(M);
@ -8520,8 +8512,8 @@ public:
L.Components.back().isNonContiguous();
generateInfoForComponentList(
L.MapType, L.MapModifiers, L.MotionModifiers, L.Components, CurInfo,
PartialStruct, IsFirstComponentList, L.IsImplicit, L.Mapper,
L.ForDeviceAddr, VD, L.VarRef);
PartialStruct, /*IsFirstComponentList=*/false, L.IsImplicit,
L.Mapper, L.ForDeviceAddr, VD, L.VarRef);
// If this entry relates with a device pointer, set the relevant
// declaration and add the 'return pointer' flag.
@ -8538,7 +8530,6 @@ public:
RelevantVD);
CurInfo.Types[CurrentBasePointersIdx] |= OMP_MAP_RETURN_PARAM;
}
IsFirstComponentList = false;
}
// Append any pending zero-length pointers which are struct members and
@ -8580,8 +8571,7 @@ public:
// If there is an entry in PartialStruct it means we have a struct with
// individual members mapped. Emit an extra combined entry.
if (PartialStruct.Base.isValid())
emitCombinedEntry(CombinedInfo, CurInfo.Types, PartialStruct, VD,
NotTargetParams);
emitCombinedEntry(CombinedInfo, CurInfo.Types, PartialStruct, VD);
// We need to append the results of this capture to what we already have.
CombinedInfo.append(CurInfo);
@ -10132,7 +10122,8 @@ void CGOpenMPRuntime::emitTargetCall(
// If there is an entry in PartialStruct it means we have a struct with
// individual members mapped. Emit an extra combined entry.
if (PartialStruct.Base.isValid())
MEHandler.emitCombinedEntry(CombinedInfo, CurInfo.Types, PartialStruct);
MEHandler.emitCombinedEntry(CombinedInfo, CurInfo.Types, PartialStruct,
nullptr, /*NoTargetParam=*/false);
// We need to append the results of this capture to what we already have.
CombinedInfo.append(CurInfo);
@ -10143,8 +10134,7 @@ void CGOpenMPRuntime::emitTargetCall(
CombinedInfo.Types);
// Map any list items in a map clause that were not captures because they
// weren't referenced within the construct.
MEHandler.generateAllInfo(CombinedInfo, /*NotTargetParams=*/true,
MappedVarSet);
MEHandler.generateAllInfo(CombinedInfo, MappedVarSet);
TargetDataInfo Info;
// Fill up the arrays and create the arguments.

30
clang/test/OpenMP/declare_mapper_codegen.cpp
View File

@ -58,25 +58,25 @@
// CK0: [[TEAMNWTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 33]
// CK0-64: [[EDSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 16]
// CK0-32: [[EDSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 8]
// CK0: [[EDTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 33]
// CK0: [[EDTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 1]
// CK0-64: [[EDNWSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 16]
// CK0-32: [[EDNWSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 8]
// CK0: [[EDNWTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 33]
// CK0: [[EDNWTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 1]
// CK0-64: [[EXDSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 16]
// CK0-32: [[EXDSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 8]
// CK0: [[EXDTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 34]
// CK0: [[EXDTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 2]
// CK0-64: [[EXDNWSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 16]
// CK0-32: [[EXDNWSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 8]
// CK0: [[EXDNWTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 34]
// CK0: [[EXDNWTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 2]
// CK0-64: [[TSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 16]
// CK0-32: [[TSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 8]
// CK0: [[TTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 33]
// CK0: [[TTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 1]
// CK0-64: [[FSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 16]
// CK0-32: [[FSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 8]
// CK0: [[FTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 34]
// CK0: [[FTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 2]
// CK0-64: [[FNWSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 16]
// CK0-32: [[FNWSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 8]
// CK0: [[FNWTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 34]
// CK0: [[FNWTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 2]
class C {
public:
@ -141,10 +141,10 @@ public:
// CK0-DAG: [[MEMBER]]
// CK0-DAG: br i1 true, label %[[LTYPE:[^,]+]], label %[[MEMBERCOM:[^,]+]]
// CK0-DAG: [[MEMBERCOM]]
// CK0-DAG: [[MEMBERCOMTYPE:%.+]] = add nuw i64 32, [[SHIPRESIZE]]
// CK0-DAG: [[MEMBERCOMTYPE:%.+]] = add nuw i64 0, [[SHIPRESIZE]]
// CK0-DAG: br label %[[LTYPE]]
// CK0-DAG: [[LTYPE]]
// CK0-DAG: [[MEMBERTYPE:%.+]] = phi i64 [ 32, %[[MEMBER]] ], [ [[MEMBERCOMTYPE]], %[[MEMBERCOM]] ]
// CK0-DAG: [[MEMBERTYPE:%.+]] = phi i64 [ 0, %[[MEMBER]] ], [ [[MEMBERCOMTYPE]], %[[MEMBERCOM]] ]
// CK0-DAG: [[TYPETF:%.+]] = and i64 [[TYPE]], 3
// CK0-DAG: [[ISALLOC:%.+]] = icmp eq i64 [[TYPETF]], 0
// CK0-DAG: br i1 [[ISALLOC]], label %[[ALLOC:[^,]+]], label %[[ALLOCELSE:[^,]+]]
@ -964,13 +964,13 @@ void foo(int a){
// CK4-64: [[TSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 16]
// CK4-32: [[TSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 8]
// PRESENT=0x1000 | TARGET_PARAM=0x20 | TO=0x1 = 0x1021
// CK4: [[TTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 [[#0x1021]]]
// PRESENT=0x1000 | TO=0x1 = 0x1001
// CK4: [[TTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 [[#0x1001]]]
// CK4-64: [[FSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 16]
// CK4-32: [[FSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 8]
// PRESENT=0x1000 | TARGET_PARAM=0x20 | FROM=0x2 = 0x1022
// CK4: [[FTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 [[#0x1022]]]
// PRESENT=0x1000 | FROM=0x2 = 0x1002
// CK4: [[FTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 [[#0x1002]]]
class C {
public:
@ -1035,10 +1035,10 @@ public:
// CK4-DAG: [[MEMBER]]
// CK4-DAG: br i1 true, label %[[LTYPE:[^,]+]], label %[[MEMBERCOM:[^,]+]]
// CK4-DAG: [[MEMBERCOM]]
// CK4-DAG: [[MEMBERCOMTYPE:%.+]] = add nuw i64 32, [[SHIPRESIZE]]
// CK4-DAG: [[MEMBERCOMTYPE:%.+]] = add nuw i64 0, [[SHIPRESIZE]]
// CK4-DAG: br label %[[LTYPE]]
// CK4-DAG: [[LTYPE]]
// CK4-DAG: [[MEMBERTYPE:%.+]] = phi i64 [ 32, %[[MEMBER]] ], [ [[MEMBERCOMTYPE]], %[[MEMBERCOM]] ]
// CK4-DAG: [[MEMBERTYPE:%.+]] = phi i64 [ 0, %[[MEMBER]] ], [ [[MEMBERCOMTYPE]], %[[MEMBERCOM]] ]
// CK4-DAG: [[TYPETF:%.+]] = and i64 [[TYPE]], 3
// CK4-DAG: [[ISALLOC:%.+]] = icmp eq i64 [[TYPETF]], 0
// CK4-DAG: br i1 [[ISALLOC]], label %[[ALLOC:[^,]+]], label %[[ALLOCELSE:[^,]+]]

52
clang/test/OpenMP/target_data_codegen.cpp
View File

@ -30,19 +30,19 @@ ST<int> gb;
double gc[100];
// CK1: [[SIZE00:@.+]] = {{.+}}constant [1 x i[[sz:64|32]]] [i{{64|32}} 800]
// CK1: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] [i64 34]
// CK1: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] [i64 2]
// CK1: [[SIZE02:@.+]] = {{.+}}constant [1 x i[[sz]]] [i[[sz]] 4]
// CK1: [[MTYPE02:@.+]] = {{.+}}constant [1 x i64] [i64 33]
// CK1: [[MTYPE02:@.+]] = {{.+}}constant [1 x i64] [i64 1]
// CK1: [[MTYPE03:@.+]] = {{.+}}constant [1 x i64] [i64 37]
// CK1: [[MTYPE03:@.+]] = {{.+}}constant [1 x i64] [i64 5]
// CK1: [[SIZE04:@.+]] = {{.+}}constant [2 x i64] [i64 sdiv exact (i64 sub (i64 ptrtoint (double** getelementptr (double*, double** getelementptr inbounds (%struct.ST, %struct.ST* @gb, i32 0, i32 1), i32 1) to i64), i64 ptrtoint (double** getelementptr inbounds (%struct.ST, %struct.ST* @gb, i32 0, i32 1) to i64)), i64 ptrtoint (i8* getelementptr (i8, i8* null, i32 1) to i64)), i64 24]
// CK1: [[MTYPE04:@.+]] = {{.+}}constant [2 x i64] [i64 32, i64 281474976710673]
// CK1: [[MTYPE04:@.+]] = {{.+}}constant [2 x i64] [i64 0, i64 281474976710673]
// CK1: [[MTYPE05:@.+]] = {{.+}}constant [1 x i64] [i64 1057]
// CK1: [[MTYPE05:@.+]] = {{.+}}constant [1 x i64] [i64 1025]
// CK1: [[MTYPE06:@.+]] = {{.+}}constant [1 x i64] [i64 1061]
// CK1: [[MTYPE06:@.+]] = {{.+}}constant [1 x i64] [i64 1029]
// CK1-LABEL: _Z3fooi
void foo(int arg) {
@ -255,17 +255,17 @@ struct ST {
ST<int> gb;
double gc[100];
// PRESENT=0x1000 | TARGET_PARAM=0x20 | TO=0x1 = 0x1021
// CK1A: [[MTYPE00Begin:@.+]] = {{.+}}constant [1 x i64] [i64 [[#0x1021]]]
// PRESENT=0x1000 | TO=0x1 = 0x1001
// CK1A: [[MTYPE00Begin:@.+]] = {{.+}}constant [1 x i64] [i64 [[#0x1001]]]
// TARGET_PARAM=0x20 | TO=0x1 = 0x21
// CK1A: [[MTYPE00End:@.+]] = {{.+}}constant [1 x i64] [i64 [[#0x21]]]
// TO=0x1 = 0x1
// CK1A: [[MTYPE00End:@.+]] = {{.+}}constant [1 x i64] [i64 [[#0x1]]]
// PRESENT=0x1000 | CLOSE=0x400 | TARGET_PARAM=0x20 | ALWAYS=0x4 | TO=0x1 = 0x1425
// CK1A: [[MTYPE01Begin:@.+]] = {{.+}}constant [1 x i64] [i64 [[#0x1425]]]
// PRESENT=0x1000 | CLOSE=0x400 | ALWAYS=0x4 | TO=0x1 = 0x1405
// CK1A: [[MTYPE01Begin:@.+]] = {{.+}}constant [1 x i64] [i64 [[#0x1405]]]
// CLOSE=0x400 | TARGET_PARAM=0x20 | ALWAYS=0x4 | TO=0x1 = 0x425
// CK1A: [[MTYPE01End:@.+]] = {{.+}}constant [1 x i64] [i64 [[#0x425]]]
// CLOSE=0x400 | ALWAYS=0x4 | TO=0x1 = 0x405
// CK1A: [[MTYPE01End:@.+]] = {{.+}}constant [1 x i64] [i64 [[#0x405]]]
// CK1A-LABEL: _Z3fooi
void foo(int arg) {
@ -357,7 +357,7 @@ struct ST {
}
};
// CK2: [[MTYPE00:@.+]] = {{.+}}constant [2 x i64] [i64 32, i64 281474976710677]
// CK2: [[MTYPE00:@.+]] = {{.+}}constant [2 x i64] [i64 0, i64 281474976710677]
// CK2-LABEL: _Z3bari
int bar(int arg){
@ -475,7 +475,7 @@ struct STT {
}
};
// CK4: [[MTYPE00:@.+]] = {{.+}}constant [2 x i64] [i64 32, i64 281474976711701]
// CK4: [[MTYPE00:@.+]] = {{.+}}constant [2 x i64] [i64 0, i64 281474976711701]
// CK4-LABEL: _Z3bari
int bar(int arg){
@ -561,7 +561,7 @@ struct S2 {
void test_close_modifier(int arg) {
S2 *ps;
// CK5: private unnamed_addr constant [5 x i64] [i64 1059, i64 32, i64 562949953421328, i64 16, i64 1043]
// CK5: private unnamed_addr constant [5 x i64] [i64 1027, i64 0, i64 562949953421328, i64 16, i64 1043]
#pragma omp target data map(close,tofrom: arg, ps->ps->ps->ps->s)
{
++(arg);
@ -585,7 +585,7 @@ void test_close_modifier(int arg) {
// SIMD-ONLY2-NOT: {{__kmpc|__tgt}}
#ifdef CK6
void test_close_modifier(int arg) {
// CK6: private unnamed_addr constant [1 x i64] [i64 1059]
// CK6: private unnamed_addr constant [1 x i64] [i64 1027]
#pragma omp target data map(close,tofrom: arg)
{++arg;}
}
@ -644,30 +644,30 @@ void test_present_modifier(int arg) {
// ps1
//
// PRESENT=0x1000 | TARGET_PARAM=0x20 = 0x1020
// PRESENT=0x1000 = 0x1000
// MEMBER_OF_1=0x1000000000000 | FROM=0x2 | TO=0x1 = 0x1000000000003
// MEMBER_OF_1=0x1000000000000 | PRESENT=0x1000 | PTR_AND_OBJ=0x10 = 0x1000000001010
// PRESENT=0x1000 | PTR_AND_OBJ=0x10 = 0x1010
// PRESENT=0x1000 | PTR_AND_OBJ=0x10 | FROM=0x2 | TO=0x1 = 0x1013
//
// CK8-SAME: {{^}} [i64 [[#0x1020]], i64 [[#0x1000000000003]],
// CK8-SAME: {{^}} [i64 [[#0x1000]], i64 [[#0x1000000000003]],
// CK8-SAME: {{^}} i64 [[#0x1000000001010]], i64 [[#0x1010]], i64 [[#0x1013]],
// arg
//
// PRESENT=0x1000 | TARGET_PARAM=0x20 | FROM=0x2 | TO=0x1 = 0x1023
// PRESENT=0x1000 | FROM=0x2 | TO=0x1 = 0x1003
//
// CK8-SAME: {{^}} i64 [[#0x1023]],
// CK8-SAME: {{^}} i64 [[#0x1003]],
// ps2
//
// PRESENT=0x1000 | TARGET_PARAM=0x20 = 0x1020
// PRESENT=0x1000 = 0x1000
// MEMBER_OF_7=0x7000000000000 | PRESENT=0x1000 | FROM=0x2 | TO=0x1 = 0x7000000001003
// MEMBER_OF_7=0x7000000000000 | PTR_AND_OBJ=0x10 = 0x7000000000010
// PTR_AND_OBJ=0x10 = 0x10
// PTR_AND_OBJ=0x10 | FROM=0x2 | TO=0x1 = 0x13
//
// CK8-SAME: {{^}} i64 [[#0x1020]], i64 [[#0x7000000001003]],
// CK8-SAME: {{^}} i64 [[#0x1000]], i64 [[#0x7000000001003]],
// CK8-SAME: {{^}} i64 [[#0x7000000000010]], i64 [[#0x10]], i64 [[#0x13]]]
#pragma omp target data map(tofrom: ps1->s) \
map(present,tofrom: arg, ps1->ps->ps->ps->s, ps2->s) \
@ -694,8 +694,8 @@ void test_present_modifier(int arg) {
// SIMD-ONLY2-NOT: {{__kmpc|__tgt}}
#ifdef CK9
void test_present_modifier(int arg) {
// PRESENT=0x1000 | TARGET_PARAM=0x20 | FROM=0x2 | TO=0x1 = 0x1023
// CK9: private unnamed_addr constant [1 x i64] [i64 [[#0x1023]]]
// PRESENT=0x1000 | FROM=0x2 | TO=0x1 = 0x1003
// CK9: private unnamed_addr constant [1 x i64] [i64 [[#0x1003]]]
#pragma omp target data map(present,tofrom: arg)
{++arg;}
}

4
clang/test/OpenMP/target_data_map_pointer_array_subscript_codegen.cpp
View File

@ -34,8 +34,8 @@ MyObject *objects;
#pragma omp end declare target
// CHECK-DAG: [[SIZES0:@.+]] = private unnamed_addr constant [1 x i64] [i64 {{8|4}}]
// CHECK-DAG: [[MAPS0:@.+]] = private unnamed_addr constant [1 x i64] [i64 49]
// CHECK-DAG: [[MAPS1:@.+]] = private unnamed_addr constant [2 x i64] [i64 32, i64 281474976710673]
// CHECK-DAG: [[MAPS0:@.+]] = private unnamed_addr constant [1 x i64] [i64 17]
// CHECK-DAG: [[MAPS1:@.+]] = private unnamed_addr constant [2 x i64] [i64 0, i64 281474976710673]
// CHECK: @main
int main(void) {
// CHECK: [[BPTR0:%.+]] = getelementptr inbounds [1 x i8*], [1 x i8*]* %{{.+}}, i32 0, i32 0

8
clang/test/OpenMP/target_data_use_device_addr_codegen.cpp
View File

@ -12,11 +12,11 @@
#define HEADER
// CHECK-DAG: [[SIZES1:@.+]] = private unnamed_addr constant [5 x i64] zeroinitializer
// 96 = 0x60 = OMP_MAP_TARGET_PARAM | OMP_MAP_RETURN_PARAM
// CHECK-DAG: [[MAPTYPES1:@.+]] = private unnamed_addr constant [5 x i64] [i64 96, i64 96, i64 96, i64 96, i64 96]
// 32 = 0x20 = OMP_MAP_TARGET_PARAM
// 64 = 0x40 = OMP_MAP_RETURN_PARAM
// CHECK-DAG: [[MAPTYPES1:@.+]] = private unnamed_addr constant [5 x i64] [i64 64, i64 64, i64 64, i64 64, i64 64]
// 0 = OMP_MAP_NONE
// 281474976710720 = 0x1000000000040 = OMP_MAP_MEMBER_OF | OMP_MAP_RETURN_PARAM
// CHECK-DAG: [[MAPTYPES2:@.+]] = private unnamed_addr constant [5 x i64] [i64 32, i64 281474976710720, i64 281474976710720, i64 281474976710720, i64 281474976710720]
// CHECK-DAG: [[MAPTYPES2:@.+]] = private unnamed_addr constant [5 x i64] [i64 0, i64 281474976710720, i64 281474976710720, i64 281474976710720, i64 281474976710720]
struct S {
int a = 0;
int *ptr = &a;

32
clang/test/OpenMP/target_data_use_device_ptr_codegen.cpp
View File

@ -22,18 +22,18 @@
double *g;
// CK1: @g ={{.*}} global double*
// CK1: [[MTYPE00:@.+]] = {{.*}}constant [2 x i64] [i64 51, i64 96]
// CK1: [[MTYPE01:@.+]] = {{.*}}constant [1 x i64] [i64 99]
// CK1: [[MTYPE03:@.+]] = {{.*}}constant [1 x i64] [i64 99]
// CK1: [[MTYPE04:@.+]] = {{.*}}constant [1 x i64] [i64 99]
// CK1: [[MTYPE05:@.+]] = {{.*}}constant [1 x i64] [i64 99]
// CK1: [[MTYPE06:@.+]] = {{.*}}constant [1 x i64] [i64 99]
// CK1: [[MTYPE07:@.+]] = {{.*}}constant [1 x i64] [i64 99]
// CK1: [[MTYPE08:@.+]] = {{.*}}constant [2 x i64] [i64 99, i64 35]
// CK1: [[MTYPE09:@.+]] = {{.*}}constant [2 x i64] [i64 99, i64 99]
// CK1: [[MTYPE10:@.+]] = {{.*}}constant [2 x i64] [i64 99, i64 99]
// CK1: [[MTYPE11:@.+]] = {{.*}}constant [2 x i64] [i64 35, i64 96]
// CK1: [[MTYPE12:@.+]] = {{.*}}constant [2 x i64] [i64 35, i64 96]
// CK1: [[MTYPE00:@.+]] = {{.*}}constant [2 x i64] [i64 19, i64 64]
// CK1: [[MTYPE01:@.+]] = {{.*}}constant [1 x i64] [i64 67]
// CK1: [[MTYPE03:@.+]] = {{.*}}constant [1 x i64] [i64 67]
// CK1: [[MTYPE04:@.+]] = {{.*}}constant [1 x i64] [i64 67]
// CK1: [[MTYPE05:@.+]] = {{.*}}constant [1 x i64] [i64 67]
// CK1: [[MTYPE06:@.+]] = {{.*}}constant [1 x i64] [i64 67]
// CK1: [[MTYPE07:@.+]] = {{.*}}constant [1 x i64] [i64 67]
// CK1: [[MTYPE08:@.+]] = {{.*}}constant [2 x i64] [i64 67, i64 3]
// CK1: [[MTYPE09:@.+]] = {{.*}}constant [2 x i64] [i64 67, i64 67]
// CK1: [[MTYPE10:@.+]] = {{.*}}constant [2 x i64] [i64 67, i64 67]
// CK1: [[MTYPE11:@.+]] = {{.*}}constant [2 x i64] [i64 3, i64 64]
// CK1: [[MTYPE12:@.+]] = {{.*}}constant [2 x i64] [i64 3, i64 64]
// CK1-LABEL: @_Z3foo
template<typename T>
@ -346,10 +346,10 @@ void bar(float *&a, int *&b) {
#ifdef CK2
// CK2: [[ST:%.+]] = type { double*, double** }
// CK2: [[MTYPE00:@.+]] = {{.*}}constant [2 x i64] [i64 32, i64 281474976710739]
// CK2: [[MTYPE01:@.+]] = {{.*}}constant [2 x i64] [i64 32, i64 281474976710739]
// CK2: [[MTYPE02:@.+]] = {{.*}}constant [3 x i64] [i64 35, i64 32, i64 562949953421392]
// CK2: [[MTYPE03:@.+]] = {{.*}}constant [3 x i64] [i64 32, i64 281474976710739, i64 281474976710736]
// CK2: [[MTYPE00:@.+]] = {{.*}}constant [2 x i64] [i64 0, i64 281474976710739]
// CK2: [[MTYPE01:@.+]] = {{.*}}constant [2 x i64] [i64 0, i64 281474976710739]
// CK2: [[MTYPE02:@.+]] = {{.*}}constant [3 x i64] [i64 3, i64 0, i64 562949953421392]
// CK2: [[MTYPE03:@.+]] = {{.*}}constant [3 x i64] [i64 0, i64 281474976710739, i64 281474976710736]
template <typename T>
struct ST {

2
clang/test/OpenMP/target_data_use_device_ptr_if_codegen.cpp
View File

@ -18,7 +18,7 @@
// SIMD-ONLY1-NOT: {{__kmpc|__tgt}}
#ifdef CK1
// CK1: [[MTYPE00:@.+]] = {{.*}}constant [1 x i64] [i64 99]
// CK1: [[MTYPE00:@.+]] = {{.*}}constant [1 x i64] [i64 67]
// CK1: [[MTYPE01:@.+]] = {{.*}}constant [1 x i64] [i64 288]
// CK1: [[MTYPE02:@.+]] = {{.*}}constant [1 x i64] [i64 288]

24
clang/test/OpenMP/target_enter_data_codegen.cpp
View File

@ -37,19 +37,19 @@ double gc[100];
// CK1-64: [[KMP_PRIVATES_T]] = type { [1 x i8*], [1 x i8*], [1 x i64] }
// CK1: [[SIZE00:@.+]] = {{.+}}constant [1 x i64] [i64 800]
// CK1: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] [i64 32]
// CK1: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] zeroinitializer
// CK1: [[SIZE02:@.+]] = {{.+}}constant [1 x i64] [i64 4]
// CK1: [[MTYPE02:@.+]] = {{.+}}constant [1 x i64] [i64 33]
// CK1: [[MTYPE02:@.+]] = {{.+}}constant [1 x i64] [i64 1]
// CK1: [[MTYPE03:@.+]] = {{.+}}constant [1 x i64] [i64 37]
// CK1: [[MTYPE03:@.+]] = {{.+}}constant [1 x i64] [i64 5]
// CK1: [[SIZE04:@.+]] = {{.+}}constant [2 x i64] [i64 sdiv exact (i64 sub (i64 ptrtoint (double** getelementptr (double*, double** getelementptr inbounds (%struct.ST, %struct.ST* @gb, i32 0, i32 1), i32 1) to i64), i64 ptrtoint (double** getelementptr inbounds (%struct.ST, %struct.ST* @gb, i32 0, i32 1) to i64)), i64 ptrtoint (i8* getelementptr (i8, i8* null, i32 1) to i64)), i64 24]
// CK1: [[MTYPE04:@.+]] = {{.+}}constant [2 x i64] [i64 32, i64 281474976710673]
// CK1: [[MTYPE04:@.+]] = {{.+}}constant [2 x i64] [i64 0, i64 281474976710673]
// CK1: [[MTYPE05:@.+]] = {{.+}}constant [1 x i64] [i64 1057]
// CK1: [[MTYPE05:@.+]] = {{.+}}constant [1 x i64] [i64 1025]
// CK1: [[MTYPE06:@.+]] = {{.+}}constant [1 x i64] [i64 1061]
// CK1: [[MTYPE06:@.+]] = {{.+}}constant [1 x i64] [i64 1029]
// CK1-LABEL: _Z3fooi
void foo(int arg) {
@ -264,11 +264,11 @@ struct ST {
ST<int> gb;
double gc[100];
// PRESENT=0x1000 | TARGET_PARAM=0x20 | TO=0x1 = 0x1021
// CK1A: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] [i64 [[#0x1021]]]
// PRESENT=0x1000 | TO=0x1 = 0x1001
// CK1A: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] [i64 [[#0x1001]]]
// PRESENT=0x1000 | CLOSE=0x400 | TARGET_PARAM=0x20 | ALWAYS=0x4 | TO=0x1 = 0x1425
// CK1A: [[MTYPE01:@.+]] = {{.+}}constant [1 x i64] [i64 [[#0x1425]]]
// PRESENT=0x1000 | CLOSE=0x400 | ALWAYS=0x4 | TO=0x1 = 0x1425
// CK1A: [[MTYPE01:@.+]] = {{.+}}constant [1 x i64] [i64 [[#0x1405]]]
// CK1A-LABEL: _Z3fooi
void foo(int arg) {
@ -354,7 +354,7 @@ struct ST {
}
};
// CK2: [[MTYPE00:@.+]] = {{.+}}constant [2 x i64] [i64 32, i64 281474976710677]
// CK2: [[MTYPE00:@.+]] = {{.+}}constant [2 x i64] [i64 0, i64 281474976710677]
// CK2-LABEL: _Z3bari
int bar(int arg){
@ -504,7 +504,7 @@ struct STT {
}
};
// CK5: [[MTYPE00:@.+]] = {{.+}}constant [2 x i64] [i64 32, i64 281474976711701]
// CK5: [[MTYPE00:@.+]] = {{.+}}constant [2 x i64] [i64 0, i64 281474976711701]
// CK5-LABEL: _Z3bari
int bar(int arg){

8
clang/test/OpenMP/target_enter_data_depend_codegen.cpp
View File

@ -30,15 +30,15 @@ ST<int> gb;
double gc[100];
// CK1: [[SIZE00:@.+]] = {{.+}}constant [1 x i64] [i64 800]
// CK1: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] [i64 32]
// CK1: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] zeroinitializer
// CK1: [[SIZE02:@.+]] = {{.+}}constant [1 x i64] [i64 4]
// CK1: [[MTYPE02:@.+]] = {{.+}}constant [1 x i64] [i64 33]
// CK1: [[MTYPE02:@.+]] = {{.+}}constant [1 x i64] [i64 1]
// CK1: [[MTYPE03:@.+]] = {{.+}}constant [1 x i64] [i64 32]
// CK1: [[MTYPE03:@.+]] = {{.+}}constant [1 x i64] zeroinitializer
// CK1: [[SIZE04:@.+]] = {{.+}}constant [2 x i64] [i64 sdiv exact (i64 sub (i64 ptrtoint (double** getelementptr (double*, double** getelementptr inbounds (%struct.ST, %struct.ST* @gb, i32 0, i32 1), i32 1) to i64), i64 ptrtoint (double** getelementptr inbounds (%struct.ST, %struct.ST* @gb, i32 0, i32 1) to i64)), i64 ptrtoint (i8* getelementptr (i8, i8* null, i32 1) to i64)), i64 24]
// CK1: [[MTYPE04:@.+]] = {{.+}}constant [2 x i64] [i64 32, i64 281474976710673]
// CK1: [[MTYPE04:@.+]] = {{.+}}constant [2 x i64] [i64 0, i64 281474976710673]
// CK1-LABEL: _Z3fooi
void foo(int arg) {

16
clang/test/OpenMP/target_exit_data_codegen.cpp
View File

@ -36,19 +36,19 @@ double gc[100];
// CK1-64: [[KMP_PRIVATES_T]] = type { [1 x i8*], [1 x i8*], [1 x i64] }
// CK1: [[SIZE00:@.+]] = {{.+}}constant [1 x i64] [i64 800]
// CK1: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] [i64 34]
// CK1: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] [i64 2]
// CK1: [[SIZE02:@.+]] = {{.+}}constant [1 x i64] [i64 4]
// CK1: [[MTYPE02:@.+]] = {{.+}}constant [1 x i64] [i64 32]
// CK1: [[MTYPE02:@.+]] = {{.+}}constant [1 x i64] zeroinitializer
// CK1: [[MTYPE03:@.+]] = {{.+}}constant [1 x i64] [i64 38]
// CK1: [[MTYPE03:@.+]] = {{.+}}constant [1 x i64] [i64 6]
// CK1: [[SIZE04:@.+]] = {{.+}}constant [2 x i64] [i64 sdiv exact (i64 sub (i64 ptrtoint (double** getelementptr (double*, double** getelementptr inbounds (%struct.ST, %struct.ST* @gb, i32 0, i32 1), i32 1) to i64), i64 ptrtoint (double** getelementptr inbounds (%struct.ST, %struct.ST* @gb, i32 0, i32 1) to i64)), i64 ptrtoint (i8* getelementptr (i8, i8* null, i32 1) to i64)), i64 24]
// CK1: [[MTYPE04:@.+]] = {{.+}}constant [2 x i64] [i64 32, i64 281474976710672]
// CK1: [[MTYPE04:@.+]] = {{.+}}constant [2 x i64] [i64 0, i64 281474976710672]
// CK1: [[MTYPE05:@.+]] = {{.+}}constant [1 x i64] [i64 1058]
// CK1: [[MTYPE05:@.+]] = {{.+}}constant [1 x i64] [i64 1026]
// CK1: [[MTYPE06:@.+]] = {{.+}}constant [1 x i64] [i64 1062]
// CK1: [[MTYPE06:@.+]] = {{.+}}constant [1 x i64] [i64 1030]
// CK1-LABEL: _Z3fooi
void foo(int arg) {
@ -265,7 +265,7 @@ struct ST {
}
};
// CK2: [[MTYPE00:@.+]] = {{.+}}constant [2 x i64] [i64 32, i64 281474976710676]
// CK2: [[MTYPE00:@.+]] = {{.+}}constant [2 x i64] [i64 0, i64 281474976710676]
// CK2-LABEL: _Z3bari
int bar(int arg){
@ -369,7 +369,7 @@ struct STT {
}
};
// CK4: [[MTYPE00:@.+]] = {{.+}}constant [2 x i64] [i64 32, i64 281474976711700]
// CK4: [[MTYPE00:@.+]] = {{.+}}constant [2 x i64] [i64 0, i64 281474976711700]
// CK4-LABEL: _Z3bari
int bar(int arg){

8
clang/test/OpenMP/target_exit_data_depend_codegen.cpp
View File

@ -30,15 +30,15 @@ ST<int> gb;
double gc[100];
// CK1: [[SIZE00:@.+]] = {{.+}}constant [1 x i64] [i64 800]
// CK1: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] [i64 34]
// CK1: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] [i64 2]
// CK1: [[SIZE02:@.+]] = {{.+}}constant [1 x i64] [i64 4]
// CK1: [[MTYPE02:@.+]] = {{.+}}constant [1 x i64] [i64 40]
// CK1: [[MTYPE02:@.+]] = {{.+}}constant [1 x i64] [i64 8]
// CK1: [[MTYPE03:@.+]] = {{.+}}constant [1 x i64] [i64 34]
// CK1: [[MTYPE03:@.+]] = {{.+}}constant [1 x i64] [i64 2]
// CK1: [[SIZE04:@.+]] = {{.+}}constant [2 x i64] [i64 sdiv exact (i64 sub (i64 ptrtoint (double** getelementptr (double*, double** getelementptr inbounds (%struct.ST, %struct.ST* @gb, i32 0, i32 1), i32 1) to i64), i64 ptrtoint (double** getelementptr inbounds (%struct.ST, %struct.ST* @gb, i32 0, i32 1) to i64)), i64 ptrtoint (i8* getelementptr (i8, i8* null, i32 1) to i64)), i64 24]
// CK1: [[MTYPE04:@.+]] = {{.+}}constant [2 x i64] [i64 32, i64 281474976710674]
// CK1: [[MTYPE04:@.+]] = {{.+}}constant [2 x i64] [i64 0, i64 281474976710674]
// CK1-LABEL: _Z3fooi
void foo(int arg) {

6
clang/test/OpenMP/target_map_member_expr_array_section_codegen.cpp
View File

@ -5,11 +5,11 @@
#ifndef HEADER
#define HEADER
// 32 = 0x20 = OMP_MAP_TARGET_PARAM
// 0 = OMP_MAP_NONE
// 281474976710656 = 0x1000000000000 = OMP_MAP_MEMBER_OF of 1-st element
// CHECK: [[MAP_ENTER:@.+]] = private unnamed_addr constant [2 x i64] [i64 32, i64 281474976710656]
// CHECK: [[MAP_ENTER:@.+]] = private unnamed_addr constant [2 x i64] [i64 0, i64 281474976710656]
// 281474976710664 = 0x1000000000008 = OMP_MAP_MEMBER_OF of 1-st element | OMP_MAP_DELETE
// CHECK: [[MAP_EXIT:@.+]] = private unnamed_addr constant [2 x i64] [i64 32, i64 281474976710664]
// CHECK: [[MAP_EXIT:@.+]] = private unnamed_addr constant [2 x i64] [i64 0, i64 281474976710664]
template <typename T>
struct S {
constexpr static int size = 6;

60
clang/test/OpenMP/target_update_codegen.cpp
View File

@ -36,15 +36,15 @@ double gc[100];
// CK1-64: [[KMP_PRIVATES_T]] = type { [1 x i8*], [1 x i8*], [1 x i64] }
// CK1: [[SIZE00:@.+]] = {{.+}}constant [1 x i[[sz:64|32]]] [i{{64|32}} 800]
// CK1: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] [i64 34]
// CK1: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] [i64 2]
// CK1: [[SIZE02:@.+]] = {{.+}}constant [1 x i[[sz]]] [i[[sz]] 4]
// CK1: [[MTYPE02:@.+]] = {{.+}}constant [1 x i64] [i64 33]
// CK1: [[MTYPE02:@.+]] = {{.+}}constant [1 x i64] [i64 1]
// CK1: [[MTYPE03:@.+]] = {{.+}}constant [1 x i64] [i64 34]
// CK1: [[MTYPE03:@.+]] = {{.+}}constant [1 x i64] [i64 2]
// CK1: [[SIZE04:@.+]] = {{.+}}constant [2 x i64] [i64 sdiv exact (i64 sub (i64 ptrtoint (double** getelementptr (double*, double** getelementptr inbounds (%struct.ST, %struct.ST* @gb, i32 0, i32 1), i32 1) to i64), i64 ptrtoint (double** getelementptr inbounds (%struct.ST, %struct.ST* @gb, i32 0, i32 1) to i64)), i64 ptrtoint (i8* getelementptr (i8, i8* null, i32 1) to i64)), i64 24]
// CK1: [[MTYPE04:@.+]] = {{.+}}constant [2 x i64] [i64 32, i64 281474976710673]
// CK1: [[MTYPE04:@.+]] = {{.+}}constant [2 x i64] [i64 0, i64 281474976710673]
// CK1-LABEL: _Z3fooi
void foo(int arg) {
@ -205,7 +205,7 @@ struct ST {
}
};
// CK2: [[MTYPE00:@.+]] = {{.+}}constant [2 x i64] [i64 32, i64 281474976710674]
// CK2: [[MTYPE00:@.+]] = {{.+}}constant [2 x i64] [i64 0, i64 281474976710674]
// CK2-LABEL: _Z3bari
int bar(int arg){
@ -342,7 +342,7 @@ void device_side_scan(int arg) {
#ifdef CK5
// CK5: [[SIZE00:@.+]] = {{.+}}constant [1 x i[[sz:64|32]]] [i{{64|32}} 4]
// CK5: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] [i64 33]
// CK5: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] [i64 1]
// CK5-LABEL: lvalue
void lvalue(int *B, int l, int e) {
@ -384,7 +384,7 @@ void lvalue(int *B, int l, int e) {
#ifdef CK6
// CK6: [[SIZE00:@.+]] = {{.+}}constant [1 x i[[sz:64|32]]] [i{{64|32}} 4]
// CK6: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] [i64 33]
// CK6: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] [i64 1]
// CK6-LABEL: lvalue
void lvalue(int *B, int l, int e) {
@ -431,7 +431,7 @@ void lvalue(int *B, int l, int e) {
#ifdef CK7
// CK7: [[SIZE00:@.+]] = {{.+}}constant [1 x i[[sz:64|32]]] [i{{64|32}} 4]
// CK7: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] [i64 33]
// CK7: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] [i64 1]
// CK7-LABEL: lvalue
void lvalue(int *B, int l, int e) {
@ -481,7 +481,7 @@ void lvalue(int *B, int l, int e) {
#ifdef CK8
// CK8: [[SIZE00:@.+]] = {{.+}}constant [2 x i[[sz:64|32]]] [i{{64|32}} {{8|4}}, i{{64|32}} 4]
// CK8: [[MTYPE00:@.+]] = {{.+}}constant [2 x i64] [i64 33, i64 17]
// CK8: [[MTYPE00:@.+]] = {{.+}}constant [2 x i64] [i64 1, i64 17]
// CK8-LABEL: lvalue
void lvalue(int **B, int l, int e) {
@ -535,7 +535,7 @@ struct S {
double *p;
};
// CK9: [[MTYPE00:@.+]] = {{.+}}constant [2 x i64] [i64 32, i64 281474976710673]
// CK9: [[MTYPE00:@.+]] = {{.+}}constant [2 x i64] [i64 0, i64 281474976710673]
// CK9-LABEL: lvalue
void lvalue(struct S *s, int l, int e) {
@ -585,7 +585,7 @@ struct S {
double *p;
};
// CK10: [[MTYPE00:@.+]] = {{.+}}constant [2 x i64] [i64 32, i64 281474976710673]
// CK10: [[MTYPE00:@.+]] = {{.+}}constant [2 x i64] [i64 0, i64 281474976710673]
// CK10-LABEL: lvalue
void lvalue(struct S *s, int l, int e) {
@ -635,7 +635,7 @@ void lvalue(struct S *s, int l, int e) {
struct S {
double *p;
};
// CK11: [[MTYPE00:@.+]] = {{.+}}constant [2 x i64] [i64 32, i64 281474976710673]
// CK11: [[MTYPE00:@.+]] = {{.+}}constant [2 x i64] [i64 0, i64 281474976710673]
// CK11-LABEL: lvalue
void lvalue(struct S *s, int l, int e) {
@ -687,7 +687,7 @@ struct S {
double *p;
struct S *sp;
};
// CK12: [[MTYPE00:@.+]] = {{.+}}constant [3 x i64] [i64 32, i64 281474976710672, i64 17]
// CK12: [[MTYPE00:@.+]] = {{.+}}constant [3 x i64] [i64 0, i64 281474976710672, i64 17]
// CK12-LABEL: lvalue
void lvalue(struct S *s, int l, int e) {
@ -749,7 +749,7 @@ void lvalue(struct S *s, int l, int e) {
#ifdef CK13
// CK13: [[SIZE00:@.+]] = {{.+}}constant [1 x i64] [i64 4]
// CK13: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] [i64 33]
// CK13: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] [i64 1]
// CK13-LABEL: lvalue
void lvalue(int **BB, int a, int b) {
@ -796,7 +796,7 @@ void lvalue(int **BB, int a, int b) {
// SIMD-ONLY0-NOT: {{__kmpc|__tgt}}
#ifdef CK14
// CK14: [[MTYPE00:@.+]] = private {{.*}}constant [2 x i64] [i64 32, i64 281474976710673]
// CK14: [[MTYPE00:@.+]] = private {{.*}}constant [2 x i64] [i64 0, i64 281474976710673]
struct SSA {
double *p;
@ -869,7 +869,7 @@ void lvalue_member(SSA *sap) {
// SIMD-ONLY0-NOT: {{__kmpc|__tgt}}
#ifdef CK15
// CK15: [[MTYPE00:@.+]] = {{.+}}constant [2 x i64] [i64 32, i64 281474976710673]
// CK15: [[MTYPE00:@.+]] = {{.+}}constant [2 x i64] [i64 0, i64 281474976710673]
struct SSA {
double *p;
@ -935,7 +935,7 @@ void lvalue_member(SSA *sap) {
#ifdef CK16
// CK16: [[SIZE00:@.+]] = {{.+}}constant [1 x i64] [i64 4]
// CK16: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] [i64 33]
// CK16: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] [i64 1]
//CK16-LABEL: lvalue_find_base
void lvalue_find_base(float *f, int *i) {
@ -980,7 +980,7 @@ void lvalue_find_base(float *f, int *i) {
#ifdef CK17
// CK17: [[SIZE00:@.+]] = {{.+}}constant [1 x i64] [i64 4]
// CK17: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] [i64 33]
// CK17: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] [i64 1]
struct SSA {
int i;
@ -1038,8 +1038,8 @@ void lvalue_find_base(float **f, SSA *sa) {
// SIMD-ONLY18-NOT: {{__kmpc|__tgt}}
#ifdef CK18
// CK18-DAG: [[MTYPE_TO:@.+]] = {{.+}}constant [1 x i64] [i64 33]
// CK18-DAG: [[MTYPE_FROM:@.+]] = {{.+}}constant [1 x i64] [i64 34]
// CK18-DAG: [[MTYPE_TO:@.+]] = {{.+}}constant [1 x i64] [i64 1]
// CK18-DAG: [[MTYPE_FROM:@.+]] = {{.+}}constant [1 x i64] [i64 2]
//CK18-LABEL: array_shaping
void array_shaping(float *f, int sa) {
@ -1113,11 +1113,11 @@ void array_shaping(float *f, int sa) {
// SIMD-ONLY0-NOT: {{__kmpc|__tgt}}
#ifdef CK19
// PRESENT=0x1000 | TARGET_PARAM=0x20 | TO=0x1 = 0x1021
// CK19: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] [i64 [[#0x1021]]]
// PRESENT=0x1000 | TO=0x1 = 0x1021
// CK19: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] [i64 [[#0x1001]]]
// PRESENT=0x1000 | TARGET_PARAM=0x20 | FROM=0x2 = 0x1022
// CK19: [[MTYPE01:@.+]] = {{.+}}constant [1 x i64] [i64 [[#0x1022]]]
// PRESENT=0x1000 | FROM=0x2 = 0x1002
// CK19: [[MTYPE01:@.+]] = {{.+}}constant [1 x i64] [i64 [[#0x1002]]]
// CK19-LABEL: _Z13check_presenti
void check_present(int arg) {
@ -1186,7 +1186,7 @@ struct ST {
// CK20: [[STRUCT_DESCRIPTOR:%.+]] = type { i64, i64, i64 }
// CK20: [[MSIZE:@.+]] = {{.+}}constant [1 x i64] [i64 3]
// CK20: [[MTYPE:@.+]] = {{.+}}constant [1 x i64] [i64 17592186044449]
// CK20: [[MTYPE:@.+]] = {{.+}}constant [1 x i64] [i64 17592186044417]
// CK20-LABEL: _Z3foo
void foo(int arg) {
@ -1254,7 +1254,7 @@ void foo(int arg) {
// CK21: [[STRUCT_ST:%.+]] = type { [10 x [10 x [10 x double*]]] }
// CK21: [[STRUCT_DESCRIPTOR:%.+]] = type { i64, i64, i64 }
// CK21: [[MTYPE:@.+]] = {{.+}}constant [2 x i64] [i64 32, i64 299067162755073]
// CK21: [[MTYPE:@.+]] = {{.+}}constant [2 x i64] [i64 0, i64 299067162755073]
struct ST {
double *dptr[10][10][10];
@ -1333,7 +1333,7 @@ void bar() {
// CK22: [[STRUCT_DESCRIPTOR:%.+]] = type { i64, i64, i64 }
// CK22: [[MSIZE:@.+]] = {{.+}}constant [1 x i64] [i64 4]
// CK22: [[MTYPE:@.+]] = {{.+}}constant [1 x i64] [i64 17592186044449]
// CK22: [[MTYPE:@.+]] = {{.+}}constant [1 x i64] [i64 17592186044417]
struct ST {
// CK22: _ZN2ST3fooEPA10_Pi
@ -1409,7 +1409,7 @@ void bar() {
// CK23: [[STRUCT_DESCRIPTOR:%.+]] = type { i64, i64, i64 }
// CK23: [[MSIZE:@.+]] = {{.+}}constant [1 x i64] [i64 4]
// CK23: [[MTYPE:@.+]] = {{.+}}constant [1 x i64] [i64 17592186044449]
// CK23: [[MTYPE:@.+]] = {{.+}}constant [1 x i64] [i64 17592186044417]
// CK23: foo
void foo(int arg) {
@ -1486,7 +1486,7 @@ void foo(int arg) {
// CK24: [[STRUCT_DESCRIPTOR:%.+]] = type { i64, i64, i64 }
// CK24: [[MSIZE:@.+]] = {{.+}}constant [1 x i64] [i64 4]
// CK24: [[MTYPE:@.+]] = {{.+}}constant [1 x i64] [i64 17592186044449]
// CK24: [[MTYPE:@.+]] = {{.+}}constant [1 x i64] [i64 17592186044417]
// CK24: foo
void foo(int arg) {
@ -1563,7 +1563,7 @@ void foo(int arg) {
// CK25: [[STRUCT_DESCRIPTOR:%.+]] = type { i64, i64, i64 }
// CK25: [[MSIZE:@.+]] = {{.+}}constant [3 x i64] [i64 4, i64 4, i64 3]
// CK25: [[MTYPE:@.+]] = {{.+}}constant [3 x i64] [i64 17592186044449, i64 33, i64 17592186044449]
// CK25: [[MTYPE:@.+]] = {{.+}}constant [3 x i64] [i64 17592186044417, i64 1, i64 17592186044417]
// CK25-LABEL: _Z3foo
void foo(int arg) {

8
clang/test/OpenMP/target_update_depend_codegen.cpp
View File

@ -30,15 +30,15 @@ ST<int> gb;
double gc[100];
// CK1: [[SIZE00:@.+]] = {{.+}}constant [1 x i64] [i64 800]
// CK1: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] [i64 34]
// CK1: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] [i64 2]
// CK1: [[SIZE02:@.+]] = {{.+}}constant [1 x i64] [i64 4]
// CK1: [[MTYPE02:@.+]] = {{.+}}constant [1 x i64] [i64 33]
// CK1: [[MTYPE02:@.+]] = {{.+}}constant [1 x i64] [i64 1]
// CK1: [[MTYPE03:@.+]] = {{.+}}constant [1 x i64] [i64 34]
// CK1: [[MTYPE03:@.+]] = {{.+}}constant [1 x i64] [i64 2]
// CK1: [[SIZE04:@.+]] = {{.+}}constant [2 x i64] [i64 sdiv exact (i64 sub (i64 ptrtoint (double** getelementptr (double*, double** getelementptr inbounds (%struct.ST, %struct.ST* @gb, i32 0, i32 1), i32 1) to i64), i64 ptrtoint (double** getelementptr inbounds (%struct.ST, %struct.ST* @gb, i32 0, i32 1) to i64)), i64 ptrtoint (i8* getelementptr (i8, i8* null, i32 1) to i64)), i64 24]
// CK1: [[MTYPE04:@.+]] = {{.+}}constant [2 x i64] [i64 32, i64 281474976710673]
// CK1: [[MTYPE04:@.+]] = {{.+}}constant [2 x i64] [i64 0, i64 281474976710673]
// CK1-LABEL: _Z3fooi
void foo(int arg) {