diff options
| -rw-r--r-- | openmp/libomptarget/src/omptarget.cpp | 6 | ||||
| -rw-r--r-- | openmp/libomptarget/test/mapping/pr38704.c | 41 |
2 files changed, 45 insertions, 2 deletions
diff --git a/openmp/libomptarget/src/omptarget.cpp b/openmp/libomptarget/src/omptarget.cpp index 14f45312030..447820deaaf 100644 --- a/openmp/libomptarget/src/omptarget.cpp +++ b/openmp/libomptarget/src/omptarget.cpp @@ -419,7 +419,7 @@ int target_data_end(DeviceTy &Device, int32_t arg_num, void **args_base, uintptr_t ub = (uintptr_t) HstPtrBegin + data_size; Device.ShadowMtx.lock(); for (ShadowPtrListTy::iterator it = Device.ShadowPtrMap.begin(); - it != Device.ShadowPtrMap.end(); ++it) { + it != Device.ShadowPtrMap.end();) { void **ShadowHstPtrAddr = (void**) it->first; // An STL map is sorted on its keys; use this property @@ -439,7 +439,9 @@ int target_data_end(DeviceTy &Device, int32_t arg_num, void **args_base, // If the struct is to be deallocated, remove the shadow entry. if (DelEntry) { DP("Removing shadow pointer " DPxMOD "\n", DPxPTR(ShadowHstPtrAddr)); - Device.ShadowPtrMap.erase(it); + it = Device.ShadowPtrMap.erase(it); + } else { + ++it; } } Device.ShadowMtx.unlock(); diff --git a/openmp/libomptarget/test/mapping/pr38704.c b/openmp/libomptarget/test/mapping/pr38704.c new file mode 100644 index 00000000000..d1b3cc42e2f --- /dev/null +++ b/openmp/libomptarget/test/mapping/pr38704.c @@ -0,0 +1,41 @@ +// RUN: %libomptarget-compile-run-and-check-aarch64-unknown-linux-gnu +// RUN: %libomptarget-compile-run-and-check-powerpc64-ibm-linux-gnu +// RUN: %libomptarget-compile-run-and-check-powerpc64le-ibm-linux-gnu +// RUN: %libomptarget-compile-run-and-check-x86_64-pc-linux-gnu + +// Clang 6.0 doesn't use the new map interface, undefined behavior when +// the compiler emits "old" interface code for structures. +// UNSUPPORTED: clang-6 + +#include <stdio.h> +#include <stdlib.h> + +typedef struct { + int *ptr1; + int *ptr2; +} StructWithPtrs; + +int main(int argc, char *argv[]) { + StructWithPtrs s; + s.ptr1 = malloc(sizeof(int)); + s.ptr2 = malloc(2 * sizeof(int)); + +#pragma omp target map(s.ptr1[0:1], s.ptr2[0:2]) + { + s.ptr1[0] = 1; + s.ptr2[0] = 2; + s.ptr2[1] = 3; + } + + // CHECK: s.ptr1[0] = 1 + // CHECK: s.ptr2[0] = 2 + // CHECK: s.ptr2[1] = 3 + printf("s.ptr1[0] = %d\n", s.ptr1[0]); + printf("s.ptr2[0] = %d\n", s.ptr2[0]); + printf("s.ptr2[1] = %d\n", s.ptr2[1]); + + free(s.ptr1); + free(s.ptr2); + + return 0; +} |
