This revision was landed with ongoing or failed builds. This revision was automatically updated to reflect the committed changes. Closed by commit rGa419ec4f256d: Fix runtime crash inside __kmpc_init_allocator (authored by jyu2).
Repository: rG LLVM Github Monorepo CHANGES SINCE LAST ACTION https://reviews.llvm.org/D151576/new/ 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