diff options
Diffstat (limited to 'test/OpenMP/target_exit_data_codegen.cpp')
-rw-r--r-- | test/OpenMP/target_exit_data_codegen.cpp | 28 |
1 files changed, 15 insertions, 13 deletions
diff --git a/test/OpenMP/target_exit_data_codegen.cpp b/test/OpenMP/target_exit_data_codegen.cpp index a82c1c6c4aac..9359e3be451b 100644 --- a/test/OpenMP/target_exit_data_codegen.cpp +++ b/test/OpenMP/target_exit_data_codegen.cpp @@ -22,15 +22,15 @@ ST<int> gb; double gc[100]; // CK1: [[SIZE00:@.+]] = {{.+}}constant [1 x i[[sz:64|32]]] [i{{64|32}} 800] -// CK1: [[MTYPE00:@.+]] = {{.+}}constant [1 x i32] [i32 34] +// CK1: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] [i64 34] // CK1: [[SIZE02:@.+]] = {{.+}}constant [1 x i[[sz]]] [i[[sz]] 4] -// CK1: [[MTYPE02:@.+]] = {{.+}}constant [1 x i32] [i32 32] +// CK1: [[MTYPE02:@.+]] = {{.+}}constant [1 x i64] [i64 32] -// CK1: [[MTYPE03:@.+]] = {{.+}}constant [1 x i32] [i32 38] +// CK1: [[MTYPE03:@.+]] = {{.+}}constant [1 x i64] [i64 38] // CK1: [[SIZE04:@.+]] = {{.+}}constant [2 x i[[sz]]] [i[[sz]] {{8|4}}, i[[sz]] 24] -// CK1: [[MTYPE04:@.+]] = {{.+}}constant [2 x i32] [i32 32, i32 16] +// CK1: [[MTYPE04:@.+]] = {{.+}}constant [2 x i64] [i64 32, i64 16] // CK1-LABEL: _Z3fooi void foo(int arg) { @@ -39,8 +39,9 @@ void foo(int arg) { // Region 00 // CK1-NOT: __tgt_target_data_begin - // CK1-DAG: call void @__tgt_target_data_end(i32 [[DEV:%[^,]+]], i32 1, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], {{.+}}getelementptr {{.+}}[1 x i{{.+}}]* [[SIZE00]], {{.+}}getelementptr {{.+}}[1 x i{{.+}}]* [[MTYPE00]]{{.+}}) - // CK1-DAG: [[DEV]] = load i32, i32* %{{[^,]+}}, + // CK1-DAG: call void @__tgt_target_data_end_nowait(i64 [[DEV:%[^,]+]], i32 1, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], {{.+}}getelementptr {{.+}}[1 x i{{.+}}]* [[SIZE00]], {{.+}}getelementptr {{.+}}[1 x i{{.+}}]* [[MTYPE00]]{{.+}}) + // CK1-DAG: [[DEV]] = sext i32 [[DEVi32:%[^,]+]] to i64 + // CK1-DAG: [[DEVi32]] = load i32, i32* %{{[^,]+}}, // CK1-DAG: [[GEPBP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]] // CK1-DAG: [[GEPP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]] @@ -52,7 +53,7 @@ void foo(int arg) { // CK1-DAG: store [100 x double]* @gc, [100 x double]** [[PC0]] // CK1: %{{.+}} = add nsw i32 %{{[^,]+}}, 1 - #pragma omp target exit data if(1+3-5) device(arg) map(from: gc) + #pragma omp target exit data if(1+3-5) device(arg) map(from: gc) nowait {++arg;} // Region 01 @@ -64,7 +65,7 @@ void foo(int arg) { // CK1-NOT: __tgt_target_data_begin // CK1: br i1 %{{[^,]+}}, label %[[IFTHEN:[^,]+]], label %[[IFELSE:[^,]+]] // CK1: [[IFTHEN]] - // CK1-DAG: call void @__tgt_target_data_end(i32 4, i32 1, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], {{.+}}getelementptr {{.+}}[1 x i{{.+}}]* [[SIZE02]], {{.+}}getelementptr {{.+}}[1 x i{{.+}}]* [[MTYPE02]]{{.+}}) + // CK1-DAG: call void @__tgt_target_data_end(i64 4, i32 1, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], {{.+}}getelementptr {{.+}}[1 x i{{.+}}]* [[SIZE02]], {{.+}}getelementptr {{.+}}[1 x i{{.+}}]* [[MTYPE02]]{{.+}}) // CK1-DAG: [[GEPBP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]] // CK1-DAG: [[GEPP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]] @@ -88,7 +89,7 @@ void foo(int arg) { // Region 03 // CK1-NOT: __tgt_target_data_begin - // CK1-DAG: call void @__tgt_target_data_end(i32 -1, i32 1, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], i[[sz]]* [[GEPS:%.+]], {{.+}}getelementptr {{.+}}[1 x i{{.+}}]* [[MTYPE03]]{{.+}}) + // CK1-DAG: call void @__tgt_target_data_end(i64 -1, i32 1, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], i[[sz]]* [[GEPS:%.+]], {{.+}}getelementptr {{.+}}[1 x i{{.+}}]* [[MTYPE03]]{{.+}}) // CK1-DAG: [[GEPBP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]] // CK1-DAG: [[GEPP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]] // CK1-DAG: [[GEPS]] = getelementptr inbounds {{.+}}[[S:%[^,]+]] @@ -111,7 +112,7 @@ void foo(int arg) { // Region 04 // CK1-NOT: __tgt_target_data_begin - // CK1-DAG: call void @__tgt_target_data_end(i32 -1, i32 2, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], {{.+}}getelementptr {{.+}}[2 x i{{.+}}]* [[SIZE04]], {{.+}}getelementptr {{.+}}[2 x i{{.+}}]* [[MTYPE04]]{{.+}}) + // CK1-DAG: call void @__tgt_target_data_end(i64 -1, i32 2, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], {{.+}}getelementptr {{.+}}[2 x i{{.+}}]* [[SIZE04]], {{.+}}getelementptr {{.+}}[2 x i{{.+}}]* [[MTYPE04]]{{.+}}) // CK1-DAG: [[GEPBP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]] // CK1-DAG: [[GEPP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]] @@ -161,7 +162,7 @@ struct ST { }; // CK2: [[SIZE00:@.+]] = {{.+}}constant [2 x i[[sz:64|32]]] [i{{64|32}} {{8|4}}, i{{64|32}} 24] -// CK2: [[MTYPE00:@.+]] = {{.+}}constant [2 x i32] [i32 36, i32 20] +// CK2: [[MTYPE00:@.+]] = {{.+}}constant [2 x i64] [i64 36, i64 20] // CK2-LABEL: _Z3bari int bar(int arg){ @@ -173,8 +174,9 @@ int bar(int arg){ // CK2-NOT: __tgt_target_data_begin // CK2: br i1 %{{[^,]+}}, label %[[IFTHEN:[^,]+]], label %[[IFELSE:[^,]+]] // CK2: [[IFTHEN]] -// CK2-DAG: call void @__tgt_target_data_end(i32 [[DEV:%[^,]+]], i32 2, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], {{.+}}getelementptr {{.+}}[2 x i{{.+}}]* [[SIZE00]], {{.+}}getelementptr {{.+}}[2 x i{{.+}}]* [[MTYPE00]]{{.+}}) -// CK2-DAG: [[DEV]] = load i32, i32* %{{[^,]+}}, +// CK2-DAG: call void @__tgt_target_data_end(i64 [[DEV:%[^,]+]], i32 2, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], {{.+}}getelementptr {{.+}}[2 x i{{.+}}]* [[SIZE00]], {{.+}}getelementptr {{.+}}[2 x i{{.+}}]* [[MTYPE00]]{{.+}}) +// CK2-DAG: [[DEV]] = sext i32 [[DEVi32:%[^,]+]] to i64 +// CK2-DAG: [[DEVi32]] = load i32, i32* %{{[^,]+}}, // CK2-DAG: [[GEPBP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]] // CK2-DAG: [[GEPP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]] |