[Openmp-commits] [PATCH] D105812: [libomptarget] Update device pointer only if needed

Joel E. Denny via Phabricator via Openmp-commits openmp-commits at lists.llvm.org
Tue Jul 13 06:37:13 PDT 2021


jdenny added inline comments.


================
Comment at: openmp/libomptarget/test/mapping/device_ptr_update.c:20
+#pragma omp target enter data map(alloc:s1.p[0:10])
+
+  // DEBUG-NOT: Update pointer ([[DEV_PTR]]) -> {{\[}}[[DEV_OBJ_A]]{{\]}}
----------------
grokos wrote:
> jdenny wrote:
> > Assume we insert `s1.p = B` here.  Without this patch, the target directive below then changes `s1.p` on the device, but it doesn't with this patch, right?
> > 
> > However, I don't think OpenMP permits that line to be inserted in the case of C/C++:
> > 
> > > A pointer that has a corresponding attached pointer must not be modified for the duration of the lifetime of the list item to which the corresponding pointer is attached in the device data environment.
> > 
> > This patch can then be considered as purely an optimization, right?
> Right. Since the pointer on the host is not allowed to be reassigned to another object while the PTR_AND_OBJ pair is mapped on the device, we can take advantage of this restriction and optimize away all redundant device pointer updates.
Thanks.


Repository:
  rG LLVM Github Monorepo

CHANGES SINCE LAST ACTION
  https://reviews.llvm.org/D105812/new/

https://reviews.llvm.org/D105812



More information about the Openmp-commits mailing list