Author: abataev
Date: Thu Jan 4 12:50:08 2018
New Revision: 321820
URL: http://llvm.org/viewvc/llvm-project?rev=321820&view=rev
Log:
[OPENMP] Fix capturing of expressions in clauses.
Patch fixes incorrect capturing of the expressions in clauses with
expressions that must be captured for the combined constructs. Incorrect
capturing may lead to compiler crash during codegen phase.
Modified:
cfe/trunk/lib/Sema/SemaOpenMP.cpp
cfe/trunk/test/OpenMP/target_teams_distribute_codegen.cpp
Modified: cfe/trunk/lib/Sema/SemaOpenMP.cpp
URL:
http://llvm.org/viewvc/llvm-project/cfe/trunk/lib/Sema/SemaOpenMP.cpp?rev=321820&r1=321819&r2=321820&view=diff
==============================================================================
--- cfe/trunk/lib/Sema/SemaOpenMP.cpp (original)
+++ cfe/trunk/lib/Sema/SemaOpenMP.cpp Thu Jan 4 12:50:08 2018
@@ -7732,12 +7732,12 @@ static OpenMPDirectiveKind getOpenMPCapt
case OMPD_target_parallel:
case OMPD_target_parallel_for:
case OMPD_target_parallel_for_simd:
- case OMPD_target_teams_distribute_parallel_for:
- case OMPD_target_teams_distribute_parallel_for_simd:
CaptureRegion = OMPD_target;
break;
case OMPD_teams_distribute_parallel_for:
case OMPD_teams_distribute_parallel_for_simd:
+ case OMPD_target_teams_distribute_parallel_for:
+ case OMPD_target_teams_distribute_parallel_for_simd:
CaptureRegion = OMPD_teams;
break;
case OMPD_parallel:
@@ -7920,20 +7920,16 @@ static OpenMPDirectiveKind getOpenMPCapt
break;
case OMPC_schedule:
switch (DKind) {
- case OMPD_target_parallel_for:
- case OMPD_target_parallel_for_simd:
- case OMPD_target_teams_distribute_parallel_for:
- case OMPD_target_teams_distribute_parallel_for_simd:
- CaptureRegion = OMPD_target;
- break;
- case OMPD_teams_distribute_parallel_for:
- case OMPD_teams_distribute_parallel_for_simd:
- CaptureRegion = OMPD_teams;
- break;
case OMPD_parallel_for:
case OMPD_parallel_for_simd:
case OMPD_distribute_parallel_for:
case OMPD_distribute_parallel_for_simd:
+ case OMPD_teams_distribute_parallel_for:
+ case OMPD_teams_distribute_parallel_for_simd:
+ case OMPD_target_parallel_for:
+ case OMPD_target_parallel_for_simd:
+ case OMPD_target_teams_distribute_parallel_for:
+ case OMPD_target_teams_distribute_parallel_for_simd:
CaptureRegion = OMPD_parallel;
break;
case OMPD_for:
@@ -7991,18 +7987,14 @@ static OpenMPDirectiveKind getOpenMPCapt
case OMPD_teams_distribute_parallel_for_simd:
case OMPD_teams_distribute:
case OMPD_teams_distribute_simd:
- CaptureRegion = OMPD_teams;
- break;
case OMPD_target_teams_distribute_parallel_for:
case OMPD_target_teams_distribute_parallel_for_simd:
case OMPD_target_teams_distribute:
case OMPD_target_teams_distribute_simd:
- CaptureRegion = OMPD_target;
+ CaptureRegion = OMPD_teams;
break;
case OMPD_distribute_parallel_for:
case OMPD_distribute_parallel_for_simd:
- CaptureRegion = OMPD_parallel;
- break;
case OMPD_distribute:
case OMPD_distribute_simd:
// Do not capture thread_limit-clause expressions.
Modified: cfe/trunk/test/OpenMP/target_teams_distribute_codegen.cpp
URL:
http://llvm.org/viewvc/llvm-project/cfe/trunk/test/OpenMP/target_teams_distribute_codegen.cpp?rev=321820&r1=321819&r2=321820&view=diff
==============================================================================
--- cfe/trunk/test/OpenMP/target_teams_distribute_codegen.cpp (original)
+++ cfe/trunk/test/OpenMP/target_teams_distribute_codegen.cpp Thu Jan 4
12:50:08 2018
@@ -62,7 +62,7 @@
// CHECK-DAG: [[MAPT2:@.+]] = private unnamed_addr constant [1 x i64] [i64 288]
// CHECK-DAG: [[SIZET3:@.+]] = private unnamed_addr constant [2 x i[[SZ]]]
[i[[SZ]] 4, i[[SZ]] 2]
// CHECK-DAG: [[MAPT3:@.+]] = private unnamed_addr constant [2 x i64] [i64
288, i64 288]
-// CHECK-DAG: [[MAPT4:@.+]] = private unnamed_addr constant [9 x i64] [i64
288, i64 547, i64 288, i64 547, i64 547, i64 288, i64 288, i64 547, i64 547]
+// CHECK-DAG: [[MAPT4:@.+]] = private unnamed_addr constant [10 x i64] [i64
288, i64 547, i64 288, i64 547, i64 547, i64 288, i64 288, i64 547, i64 547,
i64 288]
// CHECK-DAG: [[MAPT5:@.+]] = private unnamed_addr constant [5 x i64] [i64
547, i64 288, i64 288, i64 288, i64 547]
// CHECK-DAG: [[SIZET6:@.+]] = private unnamed_addr constant [5 x i[[SZ]]]
[i[[SZ]] 4, i[[SZ]] 4, i[[SZ]] 2, i[[SZ]] 1, i[[SZ]] 40]
// CHECK-DAG: [[MAPT6:@.+]] = private unnamed_addr constant [5 x i64] [i64
288, i64 288, i64 288, i64 288, i64 547]
@@ -211,6 +211,7 @@ int foo(int n) {
}
// We capture 3 VLA sizes in this target region
+ // CHECK: load i32, i32* %
// CHECK-64: [[A_VAL:%.+]] = load i32, i32* %{{.+}},
// CHECK-64: [[A_ADDR:%.+]] = bitcast i[[SZ]]* [[A_CADDR:%.+]] to i32*
// CHECK-64: store i32 [[A_VAL]], i32* [[A_ADDR]],
@@ -227,38 +228,41 @@ int foo(int n) {
// CHECK: [[IF:%.+]] = icmp sgt i32 {{[^,]+}}, 20
// CHECK: br i1 [[IF]], label %[[TRY:[^,]+]], label %[[FAIL:[^,]+]]
// CHECK: [[TRY]]
- // CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams(i64 -1, i8*
@{{[^,]+}}, i32 9, i8** [[BPR:%[^,]+]], i8** [[PR:%[^,]+]], i[[SZ]]*
[[SR:%[^,]+]], i64* getelementptr inbounds ([9 x i64], [9 x i64]* [[MAPT4]],
i32 0, i32 0), i32 0, i32 0)
- // CHECK-DAG: [[BPR]] = getelementptr inbounds [9 x i8*], [9 x i8*]*
[[BP:%[^,]+]], i32 0, i32 0
- // CHECK-DAG: [[PR]] = getelementptr inbounds [9 x i8*], [9 x i8*]*
[[P:%[^,]+]], i32 0, i32 0
- // CHECK-DAG: [[SR]] = getelementptr inbounds [9 x i[[SZ]]], [9 x
i[[SZ]]]* [[S:%[^,]+]], i32 0, i32 0
-
- // CHECK-DAG: [[SADDR0:%.+]] = getelementptr inbounds [9 x i[[SZ]]], [9 x
i[[SZ]]]* [[S]], i32 0, i32 [[IDX0:[0-9]+]]
- // CHECK-DAG: [[BPADDR0:%.+]] = getelementptr inbounds [9 x i8*], [9 x
i8*]* [[BP]], i32 0, i32 [[IDX0]]
- // CHECK-DAG: [[PADDR0:%.+]] = getelementptr inbounds [9 x i8*], [9 x
i8*]* [[P]], i32 0, i32 [[IDX0]]
- // CHECK-DAG: [[SADDR1:%.+]] = getelementptr inbounds [9 x i[[SZ]]], [9 x
i[[SZ]]]* [[S]], i32 0, i32 [[IDX1:[0-9]+]]
- // CHECK-DAG: [[BPADDR1:%.+]] = getelementptr inbounds [9 x i8*], [9 x
i8*]* [[BP]], i32 0, i32 [[IDX1]]
- // CHECK-DAG: [[PADDR1:%.+]] = getelementptr inbounds [9 x i8*], [9 x
i8*]* [[P]], i32 0, i32 [[IDX1]]
- // CHECK-DAG: [[SADDR2:%.+]] = getelementptr inbounds [9 x i[[SZ]]], [9 x
i[[SZ]]]* [[S]], i32 0, i32 [[IDX2:[0-9]+]]
- // CHECK-DAG: [[BPADDR2:%.+]] = getelementptr inbounds [9 x i8*], [9 x
i8*]* [[BP]], i32 0, i32 [[IDX2]]
- // CHECK-DAG: [[PADDR2:%.+]] = getelementptr inbounds [9 x i8*], [9 x
i8*]* [[P]], i32 0, i32 [[IDX2]]
- // CHECK-DAG: [[SADDR3:%.+]] = getelementptr inbounds [9 x i[[SZ]]], [9 x
i[[SZ]]]* [[S]], i32 0, i32 [[IDX3:[0-9]+]]
- // CHECK-DAG: [[BPADDR3:%.+]] = getelementptr inbounds [9 x i8*], [9 x
i8*]* [[BP]], i32 0, i32 [[IDX3]]
- // CHECK-DAG: [[PADDR3:%.+]] = getelementptr inbounds [9 x i8*], [9 x
i8*]* [[P]], i32 0, i32 [[IDX3]]
- // CHECK-DAG: [[SADDR4:%.+]] = getelementptr inbounds [9 x i[[SZ]]], [9 x
i[[SZ]]]* [[S]], i32 0, i32 [[IDX4:[0-9]+]]
- // CHECK-DAG: [[BPADDR4:%.+]] = getelementptr inbounds [9 x i8*], [9 x
i8*]* [[BP]], i32 0, i32 [[IDX4]]
- // CHECK-DAG: [[PADDR4:%.+]] = getelementptr inbounds [9 x i8*], [9 x
i8*]* [[P]], i32 0, i32 [[IDX4]]
- // CHECK-DAG: [[SADDR5:%.+]] = getelementptr inbounds [9 x i[[SZ]]], [9 x
i[[SZ]]]* [[S]], i32 0, i32 [[IDX5:[0-9]+]]
- // CHECK-DAG: [[BPADDR5:%.+]] = getelementptr inbounds [9 x i8*], [9 x
i8*]* [[BP]], i32 0, i32 [[IDX5]]
- // CHECK-DAG: [[PADDR5:%.+]] = getelementptr inbounds [9 x i8*], [9 x
i8*]* [[P]], i32 0, i32 [[IDX5]]
- // CHECK-DAG: [[SADDR6:%.+]] = getelementptr inbounds [9 x i[[SZ]]], [9 x
i[[SZ]]]* [[S]], i32 0, i32 [[IDX6:[0-9]+]]
- // CHECK-DAG: [[BPADDR6:%.+]] = getelementptr inbounds [9 x i8*], [9 x
i8*]* [[BP]], i32 0, i32 [[IDX6]]
- // CHECK-DAG: [[PADDR6:%.+]] = getelementptr inbounds [9 x i8*], [9 x
i8*]* [[P]], i32 0, i32 [[IDX6]]
- // CHECK-DAG: [[SADDR7:%.+]] = getelementptr inbounds [9 x i[[SZ]]], [9 x
i[[SZ]]]* [[S]], i32 0, i32 [[IDX7:[0-9]+]]
- // CHECK-DAG: [[BPADDR7:%.+]] = getelementptr inbounds [9 x i8*], [9 x
i8*]* [[BP]], i32 0, i32 [[IDX7]]
- // CHECK-DAG: [[PADDR7:%.+]] = getelementptr inbounds [9 x i8*], [9 x
i8*]* [[P]], i32 0, i32 [[IDX7]]
- // CHECK-DAG: [[SADDR8:%.+]] = getelementptr inbounds [9 x i[[SZ]]], [9 x
i[[SZ]]]* [[S]], i32 0, i32 [[IDX8:[0-9]+]]
- // CHECK-DAG: [[BPADDR8:%.+]] = getelementptr inbounds [9 x i8*], [9 x
i8*]* [[BP]], i32 0, i32 [[IDX8]]
- // CHECK-DAG: [[PADDR8:%.+]] = getelementptr inbounds [9 x i8*], [9 x
i8*]* [[P]], i32 0, i32 [[IDX8]]
+ // CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams(i64 -1, i8*
@{{[^,]+}}, i32 10, i8** [[BPR:%[^,]+]], i8** [[PR:%[^,]+]], i[[SZ]]*
[[SR:%[^,]+]], i64* getelementptr inbounds ([10 x i64], [10 x i64]* [[MAPT4]],
i32 0, i32 0), i32 0, i32 0)
+ // CHECK-DAG: [[BPR]] = getelementptr inbounds [10 x i8*], [10 x i8*]*
[[BP:%[^,]+]], i32 0, i32 0
+ // CHECK-DAG: [[PR]] = getelementptr inbounds [10 x i8*], [10 x i8*]*
[[P:%[^,]+]], i32 0, i32 0
+ // CHECK-DAG: [[SR]] = getelementptr inbounds [10 x i[[SZ]]], [10 x
i[[SZ]]]* [[S:%[^,]+]], i32 0, i32 0
+
+ // CHECK-DAG: [[SADDR0:%.+]] = getelementptr inbounds [10 x i[[SZ]]], [10
x i[[SZ]]]* [[S]], i32 0, i32 [[IDX0:[0-9]+]]
+ // CHECK-DAG: [[BPADDR0:%.+]] = getelementptr inbounds [10 x i8*], [10 x
i8*]* [[BP]], i32 0, i32 [[IDX0]]
+ // CHECK-DAG: [[PADDR0:%.+]] = getelementptr inbounds [10 x i8*], [10 x
i8*]* [[P]], i32 0, i32 [[IDX0]]
+ // CHECK-DAG: [[SADDR1:%.+]] = getelementptr inbounds [10 x i[[SZ]]], [10
x i[[SZ]]]* [[S]], i32 0, i32 [[IDX1:[0-9]+]]
+ // CHECK-DAG: [[BPADDR1:%.+]] = getelementptr inbounds [10 x i8*], [10 x
i8*]* [[BP]], i32 0, i32 [[IDX1]]
+ // CHECK-DAG: [[PADDR1:%.+]] = getelementptr inbounds [10 x i8*], [10 x
i8*]* [[P]], i32 0, i32 [[IDX1]]
+ // CHECK-DAG: [[SADDR2:%.+]] = getelementptr inbounds [10 x i[[SZ]]], [10
x i[[SZ]]]* [[S]], i32 0, i32 [[IDX2:[0-9]+]]
+ // CHECK-DAG: [[BPADDR2:%.+]] = getelementptr inbounds [10 x i8*], [10 x
i8*]* [[BP]], i32 0, i32 [[IDX2]]
+ // CHECK-DAG: [[PADDR2:%.+]] = getelementptr inbounds [10 x i8*], [10 x
i8*]* [[P]], i32 0, i32 [[IDX2]]
+ // CHECK-DAG: [[SADDR3:%.+]] = getelementptr inbounds [10 x i[[SZ]]], [10
x i[[SZ]]]* [[S]], i32 0, i32 [[IDX3:[0-9]+]]
+ // CHECK-DAG: [[BPADDR3:%.+]] = getelementptr inbounds [10 x i8*], [10 x
i8*]* [[BP]], i32 0, i32 [[IDX3]]
+ // CHECK-DAG: [[PADDR3:%.+]] = getelementptr inbounds [10 x i8*], [10 x
i8*]* [[P]], i32 0, i32 [[IDX3]]
+ // CHECK-DAG: [[SADDR4:%.+]] = getelementptr inbounds [10 x i[[SZ]]], [10
x i[[SZ]]]* [[S]], i32 0, i32 [[IDX4:[0-9]+]]
+ // CHECK-DAG: [[BPADDR4:%.+]] = getelementptr inbounds [10 x i8*], [10 x
i8*]* [[BP]], i32 0, i32 [[IDX4]]
+ // CHECK-DAG: [[PADDR4:%.+]] = getelementptr inbounds [10 x i8*], [10 x
i8*]* [[P]], i32 0, i32 [[IDX4]]
+ // CHECK-DAG: [[SADDR5:%.+]] = getelementptr inbounds [10 x i[[SZ]]], [10
x i[[SZ]]]* [[S]], i32 0, i32 [[IDX5:[0-9]+]]
+ // CHECK-DAG: [[BPADDR5:%.+]] = getelementptr inbounds [10 x i8*], [10 x
i8*]* [[BP]], i32 0, i32 [[IDX5]]
+ // CHECK-DAG: [[PADDR5:%.+]] = getelementptr inbounds [10 x i8*], [10 x
i8*]* [[P]], i32 0, i32 [[IDX5]]
+ // CHECK-DAG: [[SADDR6:%.+]] = getelementptr inbounds [10 x i[[SZ]]], [10
x i[[SZ]]]* [[S]], i32 0, i32 [[IDX6:[0-9]+]]
+ // CHECK-DAG: [[BPADDR6:%.+]] = getelementptr inbounds [10 x i8*], [10 x
i8*]* [[BP]], i32 0, i32 [[IDX6]]
+ // CHECK-DAG: [[PADDR6:%.+]] = getelementptr inbounds [10 x i8*], [10 x
i8*]* [[P]], i32 0, i32 [[IDX6]]
+ // CHECK-DAG: [[SADDR7:%.+]] = getelementptr inbounds [10 x i[[SZ]]], [10
x i[[SZ]]]* [[S]], i32 0, i32 [[IDX7:[0-9]+]]
+ // CHECK-DAG: [[BPADDR7:%.+]] = getelementptr inbounds [10 x i8*], [10 x
i8*]* [[BP]], i32 0, i32 [[IDX7]]
+ // CHECK-DAG: [[PADDR7:%.+]] = getelementptr inbounds [10 x i8*], [10 x
i8*]* [[P]], i32 0, i32 [[IDX7]]
+ // CHECK-DAG: [[SADDR8:%.+]] = getelementptr inbounds [10 x i[[SZ]]], [10
x i[[SZ]]]* [[S]], i32 0, i32 [[IDX8:[0-9]+]]
+ // CHECK-DAG: [[BPADDR8:%.+]] = getelementptr inbounds [10 x i8*], [10 x
i8*]* [[BP]], i32 0, i32 [[IDX8]]
+ // CHECK-DAG: [[PADDR8:%.+]] = getelementptr inbounds [10 x i8*], [10 x
i8*]* [[P]], i32 0, i32 [[IDX8]]
+ // CHECK-DAG: [[SADDR9:%.+]] = getelementptr inbounds [10 x i[[SZ]]], [10
x i[[SZ]]]* [[S]], i32 0, i32 [[IDX9:[0-9]+]]
+ // CHECK-DAG: [[BPADDR9:%.+]] = getelementptr inbounds [10 x i8*], [10 x
i8*]* [[BP]], i32 0, i32 [[IDX9]]
+ // CHECK-DAG: [[PADDR9:%.+]] = getelementptr inbounds [10 x i8*], [10 x
i8*]* [[P]], i32 0, i32 [[IDX9]]
// The names below are not necessarily consistent with the names used for the
// addresses above as some are repeated.
@@ -316,14 +320,20 @@ int foo(int n) {
// CHECK-DAG: [[CPADDR8]] = bitcast i8** {{%[^,]+}} to [[TT]]**
// CHECK-DAG: store i[[SZ]] {{12|16}}, i[[SZ]]* {{%[^,]+}}
+ // CHECK-DAG: store i[[SZ]] %{{.+}}, i[[SZ]]* [[CBPADDR9:%.+]],
+ // CHECK-DAG: store i[[SZ]] %{{.+}}, i[[SZ]]* [[CPADDR9:%.+]],
+ // CHECK-DAG: [[CBPADDR9]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
+ // CHECK-DAG: [[CPADDR9]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
+ // CHECK-DAG: store i[[SZ]] {{4|8}}, i[[SZ]]* {{%[^,]+}}
+
// CHECK-NEXT: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0
// CHECK-NEXT: br i1 [[ERROR]], label %[[FAIL:[^,]+]], label %[[END:[^,]+]]
// CHECK: [[FAIL]]
- // CHECK: call void [[HVT4:@.+]]({{[^,]+}}, {{[^,]+}}, {{[^,]+}},
{{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}})
+ // CHECK: call void [[HVT4:@.+]]({{[^,]+}}, {{[^,]+}}, {{[^,]+}},
{{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}})
// CHECK-NEXT: br label %[[END]]
// CHECK: [[END]]
- #pragma omp target teams distribute if(target: n>20)
+ #pragma omp target teams distribute if(target: n>20) dist_schedule(static, n)
for (int i = 0; i < 10; ++i) {
a += 1;
b[2] += 1.0;
@@ -439,6 +449,7 @@ int foo(int n) {
// CHECK: [[LOCAL_VLA3:%.+]] = alloca i[[SZ]]
// CHECK: [[LOCAL_CN:%.+]] = alloca double*
// CHECK: [[LOCAL_D:%.+]] = alloca [[TT]]*
+// CHECK: alloca i[[SZ]],
// CHECK: [[LOCAL_A_CASTED:%.+]] = alloca i[[SZ]]
// CHECK-DAG: store i[[SZ]] [[ARG_A:%.+]], i[[SZ]]* [[LOCAL_A]]
// CHECK-DAG: store [10 x float]* [[ARG_B:%.+]], [10 x float]** [[LOCAL_B]]
@@ -467,7 +478,7 @@ int foo(int n) {
// CHECK-32-DAG:store i32 [[LOCAL_AV]], i32* [[LOCAL_A_CASTED]], align
// CHECK-DAG: [[REF_A:%.+]] = load i[[SZ]], i[[SZ]]* [[LOCAL_A_CASTED]],
-// CHECK: call {{.*}}void (%ident_t*, i32, void (i32*, i32*, ...)*, ...)
@__kmpc_fork_teams(%ident_t* [[DEF_LOC]], i32 9, void (i32*, i32*, ...)*
bitcast (void (i32*, i32*, i[[SZ]], [10 x float]*, i[[SZ]], float*, [5 x [10 x
double]]*, i[[SZ]], i[[SZ]], double*, [[TT]]*)* [[OMP_OUTLINED4:@.+]] to void
(i32*, i32*, ...)*), i[[SZ]] [[REF_A]], [10 x float]* [[REF_B]], i[[SZ]]
[[VAL_VLA1]], float* [[REF_BN]], [5 x [10 x double]]* [[REF_C]], i[[SZ]]
[[VAL_VLA2]], i[[SZ]] [[VAL_VLA3]], double* [[REF_CN]], [[TT]]* [[REF_D]])
+// CHECK: call {{.*}}void (%ident_t*, i32, void (i32*, i32*, ...)*, ...)
@__kmpc_fork_teams(%ident_t* [[DEF_LOC]], i32 10, void (i32*, i32*, ...)*
bitcast (void (i32*, i32*, i[[SZ]], [10 x float]*, i[[SZ]], float*, [5 x [10 x
double]]*, i[[SZ]], i[[SZ]], double*, [[TT]]*, i[[SZ]])* [[OMP_OUTLINED4:@.+]]
to void (i32*, i32*, ...)*), i[[SZ]] [[REF_A]], [10 x float]* [[REF_B]],
i[[SZ]] [[VAL_VLA1]], float* [[REF_BN]], [5 x [10 x double]]* [[REF_C]],
i[[SZ]] [[VAL_VLA2]], i[[SZ]] [[VAL_VLA3]], double* [[REF_CN]], [[TT]]*
[[REF_D]], i[[SZ]] %{{.+}})
//
//
// CHECK: define internal {{.*}}void [[OMP_OUTLINED4]](i32* noalias
%.global_tid., i32* noalias %.bound_tid., i[[SZ]] %{{.+}}, [10 x float]*
{{.+}}, i[[SZ]] %{{.+}}, float* {{.+}}, [5 x [10 x double]]* {{.+}}, i[[SZ]]
%{{.+}}, i[[SZ]] %{{.+}}, double* {{.+}}, [[TT]]* {{.+}})
_______________________________________________
cfe-commits mailing list
cfe-commits@lists.llvm.org
http://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
