ABataev updated this revision to Diff 307694.
ABataev added a comment.
Rebase
Repository:
rG LLVM Github Monorepo
CHANGES SINCE LAST ACTION
https://reviews.llvm.org/D91261/new/
https://reviews.llvm.org/D91261
Files:
clang/lib/CodeGen/CGOpenMPRuntime.cpp
clang/test/OpenMP/declare_mapper_codegen.cpp
clang/test/OpenMP/target_data_codegen.cpp
clang/test/OpenMP/target_data_map_pointer_array_subscript_codegen.cpp
clang/test/OpenMP/target_data_use_device_addr_codegen.cpp
clang/test/OpenMP/target_data_use_device_ptr_codegen.cpp
clang/test/OpenMP/target_data_use_device_ptr_if_codegen.cpp
clang/test/OpenMP/target_enter_data_codegen.cpp
clang/test/OpenMP/target_enter_data_depend_codegen.cpp
clang/test/OpenMP/target_exit_data_codegen.cpp
clang/test/OpenMP/target_exit_data_depend_codegen.cpp
clang/test/OpenMP/target_map_member_expr_array_section_codegen.cpp
clang/test/OpenMP/target_update_codegen.cpp
clang/test/OpenMP/target_update_depend_codegen.cpp
Index: clang/test/OpenMP/target_update_depend_codegen.cpp
===================================================================
--- clang/test/OpenMP/target_update_depend_codegen.cpp
+++ clang/test/OpenMP/target_update_depend_codegen.cpp
@@ -30,15 +30,15 @@
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) {
Index: clang/test/OpenMP/target_update_codegen.cpp
===================================================================
--- clang/test/OpenMP/target_update_codegen.cpp
+++ clang/test/OpenMP/target_update_codegen.cpp
@@ -36,15 +36,15 @@
// 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 @@
}
};
-// 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 @@
#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 @@
#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 @@
#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 @@
#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 @@
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 @@
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 @@
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 @@
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 @@
#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 @@
// 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 @@
// 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 @@
#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 @@
#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 @@
// 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 @@
// 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 @@
// 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 @@
// 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 @@
// 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 @@
// 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 @@
// 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 @@
// 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) {
Index: clang/test/OpenMP/target_map_member_expr_array_section_codegen.cpp
===================================================================
--- clang/test/OpenMP/target_map_member_expr_array_section_codegen.cpp
+++ clang/test/OpenMP/target_map_member_expr_array_section_codegen.cpp
@@ -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;
Index: clang/test/OpenMP/target_exit_data_depend_codegen.cpp
===================================================================
--- clang/test/OpenMP/target_exit_data_depend_codegen.cpp
+++ clang/test/OpenMP/target_exit_data_depend_codegen.cpp
@@ -30,15 +30,15 @@
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) {
Index: clang/test/OpenMP/target_exit_data_codegen.cpp
===================================================================
--- clang/test/OpenMP/target_exit_data_codegen.cpp
+++ clang/test/OpenMP/target_exit_data_codegen.cpp
@@ -36,19 +36,19 @@
// 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 @@
}
};
-// 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 @@
}
};
-// 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){
Index: clang/test/OpenMP/target_enter_data_depend_codegen.cpp
===================================================================
--- clang/test/OpenMP/target_enter_data_depend_codegen.cpp
+++ clang/test/OpenMP/target_enter_data_depend_codegen.cpp
@@ -30,15 +30,15 @@
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) {
Index: clang/test/OpenMP/target_enter_data_codegen.cpp
===================================================================
--- clang/test/OpenMP/target_enter_data_codegen.cpp
+++ clang/test/OpenMP/target_enter_data_codegen.cpp
@@ -37,19 +37,19 @@
// 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 @@
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 @@
}
};
-// 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 @@
}
};
-// 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){
Index: clang/test/OpenMP/target_data_use_device_ptr_if_codegen.cpp
===================================================================
--- clang/test/OpenMP/target_data_use_device_ptr_if_codegen.cpp
+++ clang/test/OpenMP/target_data_use_device_ptr_if_codegen.cpp
@@ -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]
Index: clang/test/OpenMP/target_data_use_device_ptr_codegen.cpp
===================================================================
--- clang/test/OpenMP/target_data_use_device_ptr_codegen.cpp
+++ clang/test/OpenMP/target_data_use_device_ptr_codegen.cpp
@@ -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 @@
#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 {
Index: clang/test/OpenMP/target_data_use_device_addr_codegen.cpp
===================================================================
--- clang/test/OpenMP/target_data_use_device_addr_codegen.cpp
+++ clang/test/OpenMP/target_data_use_device_addr_codegen.cpp
@@ -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;
Index: clang/test/OpenMP/target_data_map_pointer_array_subscript_codegen.cpp
===================================================================
--- clang/test/OpenMP/target_data_map_pointer_array_subscript_codegen.cpp
+++ clang/test/OpenMP/target_data_map_pointer_array_subscript_codegen.cpp
@@ -34,8 +34,8 @@
#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
Index: clang/test/OpenMP/target_data_codegen.cpp
===================================================================
--- clang/test/OpenMP/target_data_codegen.cpp
+++ clang/test/OpenMP/target_data_codegen.cpp
@@ -30,19 +30,19 @@
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 @@
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 @@
}
};
-// 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 @@
}
};
-// 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 @@
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 @@
// 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 @@
// 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 @@
// 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;}
}
Index: clang/test/OpenMP/declare_mapper_codegen.cpp
===================================================================
--- clang/test/OpenMP/declare_mapper_codegen.cpp
+++ clang/test/OpenMP/declare_mapper_codegen.cpp
@@ -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 @@
// 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 @@
// 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 @@
// 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:[^,]+]]
Index: clang/lib/CodeGen/CGOpenMPRuntime.cpp
===================================================================
--- clang/lib/CodeGen/CGOpenMPRuntime.cpp
+++ clang/lib/CodeGen/CGOpenMPRuntime.cpp
@@ -8234,7 +8234,7 @@
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)
@@ -8285,7 +8285,7 @@
/// 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
@@ -8421,9 +8421,7 @@
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);
}
}
@@ -8491,19 +8489,13 @@
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);
@@ -8521,8 +8513,8 @@
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.
@@ -8539,7 +8531,6 @@
RelevantVD);
CurInfo.Types[CurrentBasePointersIdx] |= OMP_MAP_RETURN_PARAM;
}
- IsFirstComponentList = false;
}
// Append any pending zero-length pointers which are struct members and
@@ -8581,8 +8572,7 @@
// 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);
@@ -10133,7 +10123,8 @@
// 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);
@@ -10144,8 +10135,7 @@
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.
_______________________________________________
cfe-commits mailing list
cfe-commits@lists.llvm.org
https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
