[PATCH] D151576: Fix runtime crash inside __kmpc_init_allocator

Jennifer Yu via Phabricator via cfe-commits Fri, 26 May 2023 11:24:18 -0700

jyu2 created this revision.
jyu2 added reviewers: ABataev, jdoerfert.
jyu2 added projects: OpenMP, clang.
Herald added a project: All.
jyu2 requested review of this revision.
Herald added subscribers: openmp-commits, cfe-commits, jplehr, sstefan1.


It seems load of traits.addr should be passed in runtime call.  Currently
the load of load traits.addr gets passed cause runtime to fail.

To fix this, skip the call to EmitLoadOfScalar for extra load.


Repository:
  rG LLVM Github Monorepo

https://reviews.llvm.org/D151576

Files:
  clang/lib/CodeGen/CGOpenMPRuntime.cpp
  clang/test/OpenMP/target_parallel_for_simd_uses_allocators_codegen.cpp
  clang/test/OpenMP/target_parallel_for_uses_allocators_codegen.cpp
  clang/test/OpenMP/target_parallel_uses_allocators_codegen.cpp
  clang/test/OpenMP/target_simd_uses_allocators_codegen.cpp
  
clang/test/OpenMP/target_teams_distribute_parallel_for_simd_uses_allocators_codegen.cpp
  
clang/test/OpenMP/target_teams_distribute_parallel_for_uses_allocators_codegen.cpp
  clang/test/OpenMP/target_teams_distribute_simd_uses_allocators_codegen.cpp
  clang/test/OpenMP/target_teams_distribute_uses_allocators_codegen.cpp
  clang/test/OpenMP/target_teams_uses_allocators_codegen.cpp
  clang/test/OpenMP/target_uses_allocators.c
  clang/test/OpenMP/target_uses_allocators_codegen.cpp
  openmp/libomptarget/test/mapping/target_uses_allocator.c

Index: openmp/libomptarget/test/mapping/target_uses_allocator.c
===================================================================
--- /dev/null
+++ openmp/libomptarget/test/mapping/target_uses_allocator.c
@@ -0,0 +1,56 @@
+// RUN: %libomptarget-compile-run-and-check-generic
+
+#include <omp.h>
+#include <stdio.h>
+
+#define N 1024
+
+int test_omp_aligned_alloc_on_device() {
+  int errors = 0;
+
+  omp_memspace_handle_t memspace = omp_default_mem_space;
+  omp_alloctrait_t traits[2] = {{omp_atk_alignment, 64}, {omp_atk_access, 64}};
+  omp_allocator_handle_t alloc =
+      omp_init_allocator(omp_default_mem_space, 1, traits);
+
+#pragma omp target map(tofrom : errors) uses_allocators(alloc(traits))
+  {
+    int *x;
+    int not_correct_array_values = 0;
+
+    x = (int *)omp_aligned_alloc(64, N * sizeof(int), alloc);
+    if (x == NULL) {
+      errors++;
+    } else {
+#pragma omp parallel for simd simdlen(16) aligned(x : 64)
+      for (int i = 0; i < N; i++) {
+        x[i] = i;
+      }
+
+#pragma omp parallel for simd simdlen(16) aligned(x : 64)
+      for (int i = 0; i < N; i++) {
+        if (x[i] != i) {
+#pragma omp atomic write
+          not_correct_array_values = 1;
+        }
+      }
+      if (not_correct_array_values) {
+        errors++;
+      }
+      omp_free(x, alloc);
+    }
+  }
+
+  omp_destroy_allocator(alloc);
+
+  return errors;
+}
+
+int main() {
+  int errors = 0;
+  if (test_omp_aligned_alloc_on_device())
+    printf("FAILE\n");
+  else
+    // CHECK: PASSED
+    printf("PASSED\n");
+}
Index: clang/test/OpenMP/target_uses_allocators_codegen.cpp
===================================================================
--- clang/test/OpenMP/target_uses_allocators_codegen.cpp
+++ clang/test/OpenMP/target_uses_allocators_codegen.cpp
@@ -78,8 +78,7 @@
 // CHECK: [[TRAITS_ADDR_REF:%.+]] = alloca ptr,
 // CHECK: [[MY_ALLOCATOR_ADDR:%.+]] = alloca i64,
 // CHECK: [[TRAITS_ADDR:%.+]] = load ptr, ptr [[TRAITS_ADDR_REF]],
-// CHECK: [[TRAITS:%.+]] = load ptr, ptr [[TRAITS_ADDR]],
-// CHECK: [[ALLOCATOR:%.+]] = call ptr @__kmpc_init_allocator(i32 %{{.+}}, ptr null, i32 10, ptr [[TRAITS]])
+// CHECK: [[ALLOCATOR:%.+]] = call ptr @__kmpc_init_allocator(i32 %{{.+}}, ptr null, i32 10, ptr [[TRAITS_ADDR]])
 // CHECK: [[CONV:%.+]] = ptrtoint ptr [[ALLOCATOR]] to i64
 // CHECK: store i64 [[CONV]], ptr [[MY_ALLOCATOR_ADDR]],
 
Index: clang/test/OpenMP/target_uses_allocators.c
===================================================================
--- clang/test/OpenMP/target_uses_allocators.c
+++ clang/test/OpenMP/target_uses_allocators.c
@@ -132,8 +132,7 @@
 // CHECK: [[TRAITS_ADDR_REF:%.+]] = alloca ptr,
 // CHECK: [[MY_ALLOCATOR_ADDR:%.+]] = alloca i64,
 // CHECK: [[TRAITS_ADDR:%.+]] = load ptr, ptr [[TRAITS_ADDR_REF]],
-// CHECK: [[TRAITS:%.+]] = load ptr, ptr [[TRAITS_ADDR]],
-// CHECK: [[ALLOCATOR:%.+]] = call ptr @__kmpc_init_allocator(i32 %{{.+}}, ptr null, i32 10, ptr [[TRAITS]])
+// CHECK: [[ALLOCATOR:%.+]] = call ptr @__kmpc_init_allocator(i32 %{{.+}}, ptr null, i32 10, ptr [[TRAITS_ADDR]])
 // CHECK: [[CONV:%.+]] = ptrtoint ptr [[ALLOCATOR]] to i64
 // CHECK: store i64 [[CONV]], ptr [[MY_ALLOCATOR_ADDR]],
 
Index: clang/test/OpenMP/target_teams_uses_allocators_codegen.cpp
===================================================================
--- clang/test/OpenMP/target_teams_uses_allocators_codegen.cpp
+++ clang/test/OpenMP/target_teams_uses_allocators_codegen.cpp
@@ -78,8 +78,7 @@
 // CHECK: [[TRAITS_ADDR_REF:%.+]] = alloca ptr,
 // CHECK: [[MY_ALLOCATOR_ADDR:%.+]] = alloca i64,
 // CHECK: [[TRAITS_ADDR:%.+]] = load ptr, ptr [[TRAITS_ADDR_REF]],
-// CHECK: [[TRAITS:%.+]] = load ptr, ptr [[TRAITS_ADDR]],
-// CHECK: [[ALLOCATOR:%.+]] = call ptr @__kmpc_init_allocator(i32 %{{.+}}, ptr null, i32 10, ptr [[TRAITS]])
+// CHECK: [[ALLOCATOR:%.+]] = call ptr @__kmpc_init_allocator(i32 %{{.+}}, ptr null, i32 10, ptr [[TRAITS_ADDR]])
 // CHECK: [[CONV:%.+]] = ptrtoint ptr [[ALLOCATOR]] to i64
 // CHECK: store i64 [[CONV]], ptr [[MY_ALLOCATOR_ADDR]],
 
Index: clang/test/OpenMP/target_teams_distribute_uses_allocators_codegen.cpp
===================================================================
--- clang/test/OpenMP/target_teams_distribute_uses_allocators_codegen.cpp
+++ clang/test/OpenMP/target_teams_distribute_uses_allocators_codegen.cpp
@@ -79,8 +79,7 @@
 // CHECK: [[TRAITS_ADDR_REF:%.+]] = alloca ptr,
 // CHECK: [[MY_ALLOCATOR_ADDR:%.+]] = alloca i64,
 // CHECK: [[TRAITS_ADDR:%.+]] = load ptr, ptr [[TRAITS_ADDR_REF]],
-// CHECK: [[TRAITS:%.+]] = load ptr, ptr [[TRAITS_ADDR]],
-// CHECK: [[ALLOCATOR:%.+]] = call ptr @__kmpc_init_allocator(i32 %{{.+}}, ptr null, i32 10, ptr [[TRAITS]])
+// CHECK: [[ALLOCATOR:%.+]] = call ptr @__kmpc_init_allocator(i32 %{{.+}}, ptr null, i32 10, ptr [[TRAITS_ADDR]])
 // CHECK: [[CONV:%.+]] = ptrtoint ptr [[ALLOCATOR]] to i64
 // CHECK: store i64 [[CONV]], ptr [[MY_ALLOCATOR_ADDR]],
 
Index: clang/test/OpenMP/target_teams_distribute_simd_uses_allocators_codegen.cpp
===================================================================
--- clang/test/OpenMP/target_teams_distribute_simd_uses_allocators_codegen.cpp
+++ clang/test/OpenMP/target_teams_distribute_simd_uses_allocators_codegen.cpp
@@ -79,8 +79,7 @@
 // CHECK: [[TRAITS_ADDR_REF:%.+]] = alloca ptr,
 // CHECK: [[MY_ALLOCATOR_ADDR:%.+]] = alloca i64,
 // CHECK: [[TRAITS_ADDR:%.+]] = load ptr, ptr [[TRAITS_ADDR_REF]],
-// CHECK: [[TRAITS:%.+]] = load ptr, ptr [[TRAITS_ADDR]],
-// CHECK: [[ALLOCATOR:%.+]] = call ptr @__kmpc_init_allocator(i32 %{{.+}}, ptr null, i32 10, ptr [[TRAITS]])
+// CHECK: [[ALLOCATOR:%.+]] = call ptr @__kmpc_init_allocator(i32 %{{.+}}, ptr null, i32 10, ptr [[TRAITS_ADDR]])
 // CHECK: [[CONV:%.+]] = ptrtoint ptr [[ALLOCATOR]] to i64
 // CHECK: store i64 [[CONV]], ptr [[MY_ALLOCATOR_ADDR]],
 
Index: clang/test/OpenMP/target_teams_distribute_parallel_for_uses_allocators_codegen.cpp
===================================================================
--- clang/test/OpenMP/target_teams_distribute_parallel_for_uses_allocators_codegen.cpp
+++ clang/test/OpenMP/target_teams_distribute_parallel_for_uses_allocators_codegen.cpp
@@ -79,8 +79,7 @@
 // CHECK: [[TRAITS_ADDR_REF:%.+]] = alloca ptr,
 // CHECK: [[MY_ALLOCATOR_ADDR:%.+]] = alloca i64,
 // CHECK: [[TRAITS_ADDR:%.+]] = load ptr, ptr [[TRAITS_ADDR_REF]],
-// CHECK: [[TRAITS:%.+]] = load ptr, ptr [[TRAITS_ADDR]],
-// CHECK: [[ALLOCATOR:%.+]] = call ptr @__kmpc_init_allocator(i32 %{{.+}}, ptr null, i32 10, ptr [[TRAITS]])
+// CHECK: [[ALLOCATOR:%.+]] = call ptr @__kmpc_init_allocator(i32 %{{.+}}, ptr null, i32 10, ptr [[TRAITS_ADDR]])
 // CHECK: [[CONV:%.+]] = ptrtoint ptr [[ALLOCATOR]] to i64
 // CHECK: store i64 [[CONV]], ptr [[MY_ALLOCATOR_ADDR]],
 
Index: clang/test/OpenMP/target_teams_distribute_parallel_for_simd_uses_allocators_codegen.cpp
===================================================================
--- clang/test/OpenMP/target_teams_distribute_parallel_for_simd_uses_allocators_codegen.cpp
+++ clang/test/OpenMP/target_teams_distribute_parallel_for_simd_uses_allocators_codegen.cpp
@@ -79,8 +79,7 @@
 // CHECK: [[TRAITS_ADDR_REF:%.+]] = alloca ptr,
 // CHECK: [[MY_ALLOCATOR_ADDR:%.+]] = alloca i64,
 // CHECK: [[TRAITS_ADDR:%.+]] = load ptr, ptr [[TRAITS_ADDR_REF]],
-// CHECK: [[TRAITS:%.+]] = load ptr, ptr [[TRAITS_ADDR]],
-// CHECK: [[ALLOCATOR:%.+]] = call ptr @__kmpc_init_allocator(i32 %{{.+}}, ptr null, i32 10, ptr [[TRAITS]])
+// CHECK: [[ALLOCATOR:%.+]] = call ptr @__kmpc_init_allocator(i32 %{{.+}}, ptr null, i32 10, ptr [[TRAITS_ADDR]])
 // CHECK: [[CONV:%.+]] = ptrtoint ptr [[ALLOCATOR]] to i64
 // CHECK: store i64 [[CONV]], ptr [[MY_ALLOCATOR_ADDR]],
 
Index: clang/test/OpenMP/target_simd_uses_allocators_codegen.cpp
===================================================================
--- clang/test/OpenMP/target_simd_uses_allocators_codegen.cpp
+++ clang/test/OpenMP/target_simd_uses_allocators_codegen.cpp
@@ -79,8 +79,7 @@
 // CHECK: [[TRAITS_ADDR_REF:%.+]] = alloca ptr,
 // CHECK: [[MY_ALLOCATOR_ADDR:%.+]] = alloca i64,
 // CHECK: [[TRAITS_ADDR:%.+]] = load ptr, ptr [[TRAITS_ADDR_REF]],
-// CHECK: [[TRAITS:%.+]] = load ptr, ptr [[TRAITS_ADDR]],
-// CHECK: [[ALLOCATOR:%.+]] = call ptr @__kmpc_init_allocator(i32 %{{.+}}, ptr null, i32 10, ptr [[TRAITS]])
+// CHECK: [[ALLOCATOR:%.+]] = call ptr @__kmpc_init_allocator(i32 %{{.+}}, ptr null, i32 10, ptr [[TRAITS_ADDR]])
 // CHECK: [[CONV:%.+]] = ptrtoint ptr [[ALLOCATOR]] to i64
 // CHECK: store i64 [[CONV]], ptr [[MY_ALLOCATOR_ADDR]],
 
Index: clang/test/OpenMP/target_parallel_uses_allocators_codegen.cpp
===================================================================
--- clang/test/OpenMP/target_parallel_uses_allocators_codegen.cpp
+++ clang/test/OpenMP/target_parallel_uses_allocators_codegen.cpp
@@ -78,8 +78,7 @@
 // CHECK: [[TRAITS_ADDR_REF:%.+]] = alloca ptr,
 // CHECK: [[MY_ALLOCATOR_ADDR:%.+]] = alloca i64,
 // CHECK: [[TRAITS_ADDR:%.+]] = load ptr, ptr [[TRAITS_ADDR_REF]],
-// CHECK: [[TRAITS:%.+]] = load ptr, ptr [[TRAITS_ADDR]],
-// CHECK: [[ALLOCATOR:%.+]] = call ptr @__kmpc_init_allocator(i32 %{{.+}}, ptr null, i32 10, ptr [[TRAITS]])
+// CHECK: [[ALLOCATOR:%.+]] = call ptr @__kmpc_init_allocator(i32 %{{.+}}, ptr null, i32 10, ptr [[TRAITS_ADDR]])
 // CHECK: [[CONV:%.+]] = ptrtoint ptr [[ALLOCATOR]] to i64
 // CHECK: store i64 [[CONV]], ptr [[MY_ALLOCATOR_ADDR]],
 
Index: clang/test/OpenMP/target_parallel_for_uses_allocators_codegen.cpp
===================================================================
--- clang/test/OpenMP/target_parallel_for_uses_allocators_codegen.cpp
+++ clang/test/OpenMP/target_parallel_for_uses_allocators_codegen.cpp
@@ -79,8 +79,7 @@
 // CHECK: [[TRAITS_ADDR_REF:%.+]] = alloca ptr,
 // CHECK: [[MY_ALLOCATOR_ADDR:%.+]] = alloca i64,
 // CHECK: [[TRAITS_ADDR:%.+]] = load ptr, ptr [[TRAITS_ADDR_REF]],
-// CHECK: [[TRAITS:%.+]] = load ptr, ptr [[TRAITS_ADDR]],
-// CHECK: [[ALLOCATOR:%.+]] = call ptr @__kmpc_init_allocator(i32 %{{.+}}, ptr null, i32 10, ptr [[TRAITS]])
+// CHECK: [[ALLOCATOR:%.+]] = call ptr @__kmpc_init_allocator(i32 %{{.+}}, ptr null, i32 10, ptr [[TRAITS_ADDR]])
 // CHECK: [[CONV:%.+]] = ptrtoint ptr [[ALLOCATOR]] to i64
 // CHECK: store i64 [[CONV]], ptr [[MY_ALLOCATOR_ADDR]],
 
Index: clang/test/OpenMP/target_parallel_for_simd_uses_allocators_codegen.cpp
===================================================================
--- clang/test/OpenMP/target_parallel_for_simd_uses_allocators_codegen.cpp
+++ clang/test/OpenMP/target_parallel_for_simd_uses_allocators_codegen.cpp
@@ -79,8 +79,7 @@
 // CHECK: [[TRAITS_ADDR_REF:%.+]] = alloca ptr,
 // CHECK: [[MY_ALLOCATOR_ADDR:%.+]] = alloca i64,
 // CHECK: [[TRAITS_ADDR:%.+]] = load ptr, ptr [[TRAITS_ADDR_REF]],
-// CHECK: [[TRAITS:%.+]] = load ptr, ptr [[TRAITS_ADDR]],
-// CHECK: [[ALLOCATOR:%.+]] = call ptr @__kmpc_init_allocator(i32 %{{.+}}, ptr null, i32 10, ptr [[TRAITS]])
+// CHECK: [[ALLOCATOR:%.+]] = call ptr @__kmpc_init_allocator(i32 %{{.+}}, ptr null, i32 10, ptr [[TRAITS_ADDR]])
 // CHECK: [[CONV:%.+]] = ptrtoint ptr [[ALLOCATOR]] to i64
 // CHECK: store i64 [[CONV]], ptr [[MY_ALLOCATOR_ADDR]],
 
Index: clang/lib/CodeGen/CGOpenMPRuntime.cpp
===================================================================
--- clang/lib/CodeGen/CGOpenMPRuntime.cpp
+++ clang/lib/CodeGen/CGOpenMPRuntime.cpp
@@ -6034,8 +6034,7 @@
   AllocatorTraitsLVal = CGF.MakeAddrLValue(Addr, CGF.getContext().VoidPtrTy,
                                            AllocatorTraitsLVal.getBaseInfo(),
                                            AllocatorTraitsLVal.getTBAAInfo());
-  llvm::Value *Traits =
-      CGF.EmitLoadOfScalar(AllocatorTraitsLVal, AllocatorTraits->getExprLoc());
+  llvm::Value *Traits = Addr.getPointer();
 
   llvm::Value *AllocatorVal =
       CGF.EmitRuntimeCall(OMPBuilder.getOrCreateRuntimeFunction(

_______________________________________________
cfe-commits mailing list
cfe-commits@lists.llvm.org
https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits

[PATCH] D151576: Fix runtime crash inside __kmpc_init_allocator

Reply via email to