[PATCH] D91805: [OPENMP]Use the real pointer value as base, not indexed value.

[Alexey Bataev via Phabricator via cfe-commits]

ABataev created this revision.
ABataev added a reviewer: jdoerfert.
Herald added subscribers: arphaman, guansong, yaxunl.
Herald added a project: clang.
ABataev requested review of this revision.

After fix for PR48174 the base pointer for pointer-based
array-sections/array-subscripts will be emitted as `&ptr[idx]`, but
actually it should be just `ptr`, i.e. the address stored in the ponter
to point correctly to the beginning of the array. Currently it may lead
to a crash in the runtime.


Repository:
  rG LLVM Github Monorepo

https://reviews.llvm.org/D91805

Files:
  clang/lib/CodeGen/CGOpenMPRuntime.cpp
  clang/test/OpenMP/target_data_map_pointer_array_subscript_codegen.cpp


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
@@ -38,10 +38,17 @@
 // CHECK-DAG: [[MAPS1:@.+]] = private unnamed_addr constant [2 x i64] [i64 32, i64 281474976710673]
 // CHECK: @main
 int main(void) {
+// CHECK: [[BPTR0:%.+]] = getelementptr inbounds [1 x i8*], [1 x i8*]* %{{.+}}, i32 0, i32 0
+// CHECK: [[BPTRC0:%.+]] = bitcast i8** [[BPTR0]] to %struct.MyObject***
+// CHECK: store %struct.MyObject** @objects, %struct.MyObject*** [[BPTRC0]],
 // CHECK: call void @__tgt_target_data_begin_mapper(%struct.ident_t* @{{.+}}, i64 -1, i32 1, i8** %{{.+}}, i8** %{{.+}}, i64* getelementptr inbounds ([1 x i64], [1 x i64]* [[SIZES0]], i32 0, i32 0), i64* getelementptr inbounds ([1 x i64], [1 x i64]* [[MAPS0]], i32 0, i32 0), i8** null, i8** null)
-#pragma omp target enter data map(to : objects [0:1])
+#pragma omp target enter data map(to : objects [1:1])
+// CHECK: [[OBJ:%.+]] = load %struct.MyObject*, %struct.MyObject** @objects,
+// CHECK: [[BPTR0:%.+]] = getelementptr inbounds [2 x i8*], [2 x i8*]* %{{.+}}, i32 0, i32 0
+// CHECK: [[BPTRC0:%.+]] = bitcast i8** [[BPTR0]] to %struct.MyObject**
+// CHECK: store %struct.MyObject* [[OBJ]], %struct.MyObject** [[BPTRC0]],
 // CHECK: call void @__tgt_target_data_begin_mapper(%struct.ident_t* @{{.+}}, i64 -1, i32 2, i8** %{{.+}}, i8** %{{.+}}, i64* %{{.+}}, i64* getelementptr inbounds ([2 x i64], [2 x i64]* [[MAPS1]], i32 0, i32 0), i8** null, i8** null)
-#pragma omp target enter data map(to : objects[0].arr [0:1])
+#pragma omp target enter data map(to : objects[1].arr [0:1])
 
   return 0;
 }
Index: clang/lib/CodeGen/CGOpenMPRuntime.cpp
===================================================================
--- clang/lib/CodeGen/CGOpenMPRuntime.cpp
+++ clang/lib/CodeGen/CGOpenMPRuntime.cpp
@@ -7903,8 +7903,11 @@
         IsCaptureFirstInfo = false;
         FirstPointerInComplexData = false;
       } else if (FirstPointerInComplexData) {
-        BP = CGF.EmitOMPSharedLValue(I->getAssociatedExpression())
-                 .getAddress(CGF);
+        QualType Ty = Components.rbegin()
+                          ->getAssociatedDeclaration()
+                          ->getType()
+                          .getNonReferenceType();
+        BP = CGF.EmitLoadOfPointer(BP, Ty->castAs<PointerType>());
         FirstPointerInComplexData = false;
       }
     }


-------------- next part --------------
A non-text attachment was scrubbed...
Name: D91805.306449.patch
Type: text/x-patch
Size: 2570 bytes
Desc: not available
URL: <http://lists.llvm.org/pipermail/cfe-commits/attachments/20201119/35643f07/attachment.bin>