[NFC][OpenMP] Renamed all variable and function names in `target` to conform with LLVM code standard

[tianshilei1992]

Summary

This patch only touched variables and functions in target.

[ye-luo]

Should be easy to address my comments and let us get this merged ASAP.

openmp/libomptarget/src/device.h

202

"Function names should be verb phrases (as they represent actions), and command-like function should be imperative. The name should be camel case, and start with a lower case letter (e.g. openFile() or isFoo())."

Since we are anyway changing the names. Let us fix them fully.
allocData
deleteData
Please also take this chance to rename
data_retrieve
data_exchange
as well. No good reason to leave these pains.

openmp/libomptarget/src/omptarget.cpp

213

isMemberOf?

openmp/libomptarget/src/private.h

29

Let us kill the pain in one shot. It is just a few more sed.

[tianshilei1992]

Renamed some function names

[tianshilei1992]

openmp/libomptarget/src/device.h

202

I would like to just focus on the functions related to target just as the title said. I'll do it in the future if they're still there and I'm gonna touch them.

openmp/libomptarget/src/private.h

29

As replied above.

[ye-luo]

OK. Leave the unrelated renaming to the future.

openmp/libomptarget/src/api.cpp

if (device_num == omp_get_initial_device()) {
     return rc;
   }
 
   if (!device_is_ready(device_num)) {
     DP("omp_target_alloc returns NULL ptr\n");
     return NULL;
   }
 
-  rc = Devices[device_num].data_alloc(size);
+  rc = Devices[device_num].allocData(size);
   DP("omp_target_alloc returns device ptr " DPxMOD "\n", DPxPTR(rc));
   return rc;
 }
 
 EXTERN void omp_target_free(void *device_ptr, int device_num) {
   DP("Call to omp_target_free for device %d and address " DPxMOD "\n",
       device_num, DPxPTR(device_ptr));
 
   if (!device_ptr) {
     DP("Call to omp_target_free with NULL ptr\n");
     return;
   }
 
   if (device_num == omp_get_initial_device()) {
     free(device_ptr);
     DP("omp_target_free deallocated host ptr\n");
     return;
   }
 
   if (!device_is_ready(device_num)) {
     DP("omp_target_free returns, nothing to do\n");
     return;
   }
 
-  Devices[device_num].data_delete(device_ptr);
+  Devices[device_num].deleteData(device_ptr);
   DP("omp_target_free deallocated device ptr\n");
 }
 
 EXTERN int omp_target_is_present(void *ptr, int device_num) {
   DP("Call to omp_target_is_present for device %d and address " DPxMOD "\n",
       device_num, DPxPTR(ptr));
 
   if (!ptr) {
if (src_device == omp_get_initial_device() &&
       dst_device == omp_get_initial_device()) {
     DP("copy from host to host\n");
     const void *p = memcpy(dstAddr, srcAddr, length);
     if (p == NULL)
       rc = OFFLOAD_FAIL;
   } else if (src_device == omp_get_initial_device()) {
     DP("copy from host to device\n");
     DeviceTy& DstDev = Devices[dst_device];
-    rc = DstDev.data_submit(dstAddr, srcAddr, length, nullptr);
+    rc = DstDev.submitData(dstAddr, srcAddr, length, nullptr);
   } else if (dst_device == omp_get_initial_device()) {
     DP("copy from device to host\n");
     DeviceTy& SrcDev = Devices[src_device];
     rc = SrcDev.data_retrieve(dstAddr, srcAddr, length, nullptr);
   } else {
     DP("copy from device to device\n");
     DeviceTy &SrcDev = Devices[src_device];
     DeviceTy &DstDev = Devices[dst_device];
     // First try to use D2D memcpy which is more efficient. If fails, fall back
     // to unefficient way.
     if (SrcDev.isDataExchangable(DstDev)) {
       rc = SrcDev.data_exchange(srcAddr, DstDev, dstAddr, length, nullptr);
       if (rc == OFFLOAD_SUCCESS)
         return OFFLOAD_SUCCESS;
     }
 
     void *buffer = malloc(length);
     rc = SrcDev.data_retrieve(buffer, srcAddr, length, nullptr);
     if (rc == OFFLOAD_SUCCESS)
-      rc = DstDev.data_submit(dstAddr, buffer, length, nullptr);
+      rc = DstDev.submitData(dstAddr, buffer, length, nullptr);
     free(buffer);
   }
 
   DP("omp_target_memcpy returns %d\n", rc);
   return rc;
 }
 
 EXTERN int omp_target_memcpy_rect(void *dst, void *src, size_t element_size,

openmp/libomptarget/src/device.h

struct DeviceTy {
 
   // device memory allocation/deallocation routines
   /// Allocates \p Size bytes on the device and returns the address/nullptr when
   /// succeeds/fails. \p HstPtr is an address of the host data which the
   /// allocated target data will be associated with. If it is unknown, the
   /// default value of \p HstPtr is nullptr. Note: this function doesn't do
   /// pointer association. Actually, all the __tgt_rtl_data_alloc
   /// implementations ignore \p HstPtr.
-  void *data_alloc(int64_t Size, void *HstPtr = nullptr);
+  void *allocData(int64_t Size, void *HstPtr = nullptr);
   /// Deallocates memory which \p TgtPtrBegin points at and returns
   /// OFFLOAD_SUCCESS/OFFLOAD_FAIL when succeeds/fails.
-  int32_t data_delete(void *TgtPtrBegin);
+  int32_t deleteData(void *TgtPtrBegin);
 
   // Data transfer. When AsyncInfoPtr is nullptr, the transfer will be
   // synchronous.
   // Copy data from host to device
-  int32_t data_submit(void *TgtPtrBegin, void *HstPtrBegin, int64_t Size,
-                      __tgt_async_info *AsyncInfoPtr);
+  int32_t submitData(void *TgtPtrBegin, void *HstPtrBegin, int64_t Size,
   // Copy data from device back to host
   int32_t data_retrieve(void *HstPtrBegin, void *TgtPtrBegin, int64_t Size,
                         __tgt_async_info *AsyncInfoPtr);
   // Copy data from current device to destination device directly
   int32_t data_exchange(void *SrcPtr, DeviceTy DstDev, void *DstPtr,
                         int64_t Size, __tgt_async_info *AsyncInfoPtr);
 
-  int32_t run_region(void *TgtEntryPtr, void **TgtVarsPtr,
-                     ptrdiff_t *TgtOffsets, int32_t TgtVarsSize,
-                     __tgt_async_info *AsyncInfoPtr);
-  int32_t run_team_region(void *TgtEntryPtr, void **TgtVarsPtr,
-                          ptrdiff_t *TgtOffsets, int32_t TgtVarsSize,
-                          int32_t NumTeams, int32_t ThreadLimit,
-                          uint64_t LoopTripCount,
-                          __tgt_async_info *AsyncInfoPtr);
+  int32_t runRegion(void *TgtEntryPtr, void **TgtVarsPtr, ptrdiff_t *TgtOffsets,
+                    int32_t TgtVarsSize, __tgt_async_info *AsyncInfoPtr);
+  int32_t runTeamRegion(void *TgtEntryPtr, void **TgtVarsPtr,
+                        uint64_t LoopTripCount, __tgt_async_info *AsyncInfoPtr);
 
   /// Synchronize device/queue/event based on \p AsyncInfoPtr and return
   /// OFFLOAD_SUCCESS/OFFLOAD_FAIL when succeeds/fails.
   int32_t synchronize(__tgt_async_info *AsyncInfoPtr);
 
 private:
   // Call to RTL
   void init(); // To be called only via DeviceTy::initOnce()

openmp/libomptarget/src/device.cpp

LookupResult DeviceTy::lookupMapping(void *HstPtrBegin, int64_t Size) {
   if (lr.Flags.ExtendsAfter) {
     DP("WARNING: Pointer is already mapped but section extends beyond mapped "
         "region\n");
   }
 
   return lr;
 }
 
-// Used by target_data_begin
+// Used by targetDataBegin
 // Return the target pointer begin (where the data will be moved).
 // Allocate memory if this is the first occurrence of this mapping.
 // Increment the reference counter.
 // If NULL is returned, then either data allocation failed or the user tried
 // to do an illegal mapping.
 void *DeviceTy::getOrAllocTgtPtr(void *HstPtrBegin, void *HstPtrBase,
                                  int64_t Size, bool &IsNew, bool &IsHostPtr,
                                  bool IsImplicit, bool UpdateRefCount,
DP("Mapping required by 'present' map type modifier does not exist for "
        "HstPtrBegin=" DPxMOD ", Size=%ld\n",
        DPxPTR(HstPtrBegin), Size);
     MESSAGE("device mapping required by 'present' map type modifier does not "
             "exist for host address " DPxMOD " (%ld bytes)",
             DPxPTR(HstPtrBegin), Size);
   } else if (Size) {
     // If it is not contained and Size > 0, we should create a new entry for it.
     IsNew = true;
-    uintptr_t tp = (uintptr_t)data_alloc(Size, HstPtrBegin);
+    uintptr_t tp = (uintptr_t)allocData(Size, HstPtrBegin);
     DP("Creating new map entry: HstBase=" DPxMOD ", HstBegin=" DPxMOD ", "
        "HstEnd=" DPxMOD ", TgtBegin=" DPxMOD "\n",
        DPxPTR(HstPtrBase), DPxPTR(HstPtrBegin),
        DPxPTR((uintptr_t)HstPtrBegin + Size), DPxPTR(tp));
     HostDataToTargetMap.emplace(
         HostDataToTargetTy((uintptr_t)HstPtrBase, (uintptr_t)HstPtrBegin,
                            (uintptr_t)HstPtrBegin + Size, tp));
     rc = (void *)tp;
   }
 
   DataMapMtx.unlock();
   return rc;
 }
 
-// Used by target_data_begin, target_data_end, target_data_update and target.
+// Used by targetDataBegin, targetDataEnd, target_data_update and target.
 // Return the target pointer begin (where the data will be moved).
-// Decrement the reference counter if called from target_data_end.
+// Decrement the reference counter if called from targetDataEnd.
 void *DeviceTy::getTgtPtrBegin(void *HstPtrBegin, int64_t Size, bool &IsLast,
     bool UpdateRefCount, bool &IsHostPtr) {
   void *rc = NULL;
   IsHostPtr = false;
   IsLast = false;
   DataMapMtx.lock();
   LookupResult lr = lookupMapping(HstPtrBegin, Size);
 
int DeviceTy::deallocTgtPtr(void *HstPtrBegin, int64_t Size, bool ForceDelete,
   LookupResult lr = lookupMapping(HstPtrBegin, Size);
   if (lr.Flags.IsContained || lr.Flags.ExtendsBefore || lr.Flags.ExtendsAfter) {
     auto &HT = *lr.Entry;
     if (ForceDelete)
       HT.resetRefCount();
     if (HT.decRefCount() == 0) {
       DP("Deleting tgt data " DPxMOD " of size %ld\n",
           DPxPTR(HT.TgtPtrBegin), Size);
-      data_delete((void *)HT.TgtPtrBegin);
+      deleteData((void *)HT.TgtPtrBegin);
       DP("Removing%s mapping with HstPtrBegin=" DPxMOD ", TgtPtrBegin=" DPxMOD
           ", Size=%ld\n", (ForceDelete ? " (forced)" : ""),
           DPxPTR(HT.HstPtrBegin), DPxPTR(HT.TgtPtrBegin), Size);
       HostDataToTargetMap.erase(lr.Entry);
     }
     rc = OFFLOAD_SUCCESS;
   } else {
     DP("Section to delete (hst addr " DPxMOD ") does not exist in the allocated"

 // Load binary to device.
 __tgt_target_table *DeviceTy::load_binary(void *Img) {
   RTL->Mtx.lock();
   __tgt_target_table *rc = RTL->load_binary(RTLDeviceID, Img);
   RTL->Mtx.unlock();
   return rc;
 }
 
-void *DeviceTy::data_alloc(int64_t Size, void *HstPtr) {
+void *DeviceTy::allocData(int64_t Size, void *HstPtr) {
   return RTL->data_alloc(RTLDeviceID, Size, HstPtr);
 }
 
-int32_t DeviceTy::data_delete(void *TgtPtrBegin) {
+int32_t DeviceTy::deleteData(void *TgtPtrBegin) {
   return RTL->data_delete(RTLDeviceID, TgtPtrBegin);
 }
 
 // Submit data to device
-int32_t DeviceTy::data_submit(void *TgtPtrBegin, void *HstPtrBegin,
-                              int64_t Size, __tgt_async_info *AsyncInfoPtr) {
+int32_t DeviceTy::submitData(void *TgtPtrBegin, void *HstPtrBegin, int64_t Size,
+                             __tgt_async_info *AsyncInfoPtr) {
   if (!AsyncInfoPtr || !RTL->data_submit_async || !RTL->synchronize)
     return RTL->data_submit(RTLDeviceID, TgtPtrBegin, HstPtrBegin, Size);
   else
     return RTL->data_submit_async(RTLDeviceID, TgtPtrBegin, HstPtrBegin, Size,
                                   AsyncInfoPtr);
 }
 
 // Retrieve data from device
if (!AsyncInfoPtr || !RTL->data_exchange_async || !RTL->synchronize) {
     return RTL->data_exchange(RTLDeviceID, SrcPtr, DstDev.RTLDeviceID, DstPtr,
                               Size);
   } else
     return RTL->data_exchange_async(RTLDeviceID, SrcPtr, DstDev.RTLDeviceID,
                                     DstPtr, Size, AsyncInfoPtr);
 }
 
 // Run region on device
-int32_t DeviceTy::run_region(void *TgtEntryPtr, void **TgtVarsPtr,
-                             ptrdiff_t *TgtOffsets, int32_t TgtVarsSize,
-                             __tgt_async_info *AsyncInfoPtr) {
+int32_t DeviceTy::runRegion(void *TgtEntryPtr, void **TgtVarsPtr,
   if (!AsyncInfoPtr || !RTL->run_region || !RTL->synchronize)
     return RTL->run_region(RTLDeviceID, TgtEntryPtr, TgtVarsPtr, TgtOffsets,
                            TgtVarsSize);
   else
     return RTL->run_region_async(RTLDeviceID, TgtEntryPtr, TgtVarsPtr,
                                  TgtOffsets, TgtVarsSize, AsyncInfoPtr);
 }
 
 // Run team region on device.
-int32_t DeviceTy::run_team_region(void *TgtEntryPtr, void **TgtVarsPtr,
-                                  ptrdiff_t *TgtOffsets, int32_t TgtVarsSize,
-                                  int32_t NumTeams, int32_t ThreadLimit,
-                                  uint64_t LoopTripCount,
-                                  __tgt_async_info *AsyncInfoPtr) {
+int32_t DeviceTy::runTeamRegion(void *TgtEntryPtr, void **TgtVarsPtr,
   if (!AsyncInfoPtr || !RTL->run_team_region_async || !RTL->synchronize)
     return RTL->run_team_region(RTLDeviceID, TgtEntryPtr, TgtVarsPtr,
                                 TgtOffsets, TgtVarsSize, NumTeams, ThreadLimit,
                                 LoopTripCount);
   else
     return RTL->run_team_region_async(RTLDeviceID, TgtEntryPtr, TgtVarsPtr,
                                       TgtOffsets, TgtVarsSize, NumTeams,
                                       ThreadLimit, LoopTripCount, AsyncInfoPtr);

openmp/libomptarget/src/interface.cpp

 #ifdef OMPTARGET_DEBUG
   for (int i = 0; i < arg_num; ++i) {
     DP("Entry %2d: Base=" DPxMOD ", Begin=" DPxMOD ", Size=%" PRId64
        ", Type=0x%" PRIx64 "\n",
        i, DPxPTR(args_base[i]), DPxPTR(args[i]), arg_sizes[i], arg_types[i]);
   }
 #endif
 
-  int rc = target_data_begin(Device, arg_num, args_base, args, arg_sizes,
+  int rc = targetDataBegin(Device, arg_num, args_base, args, arg_sizes,
                            arg_types, arg_mappers, nullptr);
   HandleTargetOutcome(rc == OFFLOAD_SUCCESS);
 }
 
 EXTERN void __tgt_target_data_begin_nowait_mapper(int64_t device_id,
     int32_t arg_num, void **args_base, void **args, int64_t *arg_sizes,
     int64_t *arg_types, void **arg_mappers, int32_t depNum, void *depList,
     int32_t noAliasDepNum, void *noAliasDepList) {
   if (depNum + noAliasDepNum > 0)

 #ifdef OMPTARGET_DEBUG
   for (int i=0; i<arg_num; ++i) {
     DP("Entry %2d: Base=" DPxMOD ", Begin=" DPxMOD ", Size=%" PRId64
         ", Type=0x%" PRIx64 "\n", i, DPxPTR(args_base[i]), DPxPTR(args[i]),
         arg_sizes[i], arg_types[i]);
   }
 #endif
 
-  int rc = target_data_end(Device, arg_num, args_base, args, arg_sizes,
-      arg_types, arg_mappers, nullptr);
+  int rc = targetDataEnd(Device, arg_num, args_base, args, arg_sizes, arg_types,
+                         arg_mappers, nullptr);
   HandleTargetOutcome(rc == OFFLOAD_SUCCESS);
 }
 
 EXTERN void __tgt_target_data_end_nowait_mapper(int64_t device_id,
     int32_t arg_num, void **args_base, void **args, int64_t *arg_sizes,
     int64_t *arg_types, void **arg_mappers, int32_t depNum, void *depList,
     int32_t noAliasDepNum, void *noAliasDepList) {
   if (depNum + noAliasDepNum > 0)

openmp/libomptarget/src/omptarget.cpp

int CheckDeviceAndCtors(int64_t device_id) {
   if (hasPendingGlobals && InitLibrary(Device) != OFFLOAD_SUCCESS) {
     DP("Failed to init globals on device %" PRId64 "\n", device_id);
     return OFFLOAD_FAIL;
   }
 
   return OFFLOAD_SUCCESS;
 }
 
-static int32_t member_of(int64_t type) {
+static int32_t getParentIndex(int64_t type) {
   return ((type & OMP_TGT_MAPTYPE_MEMBER_OF) >> 48) - 1;
 }
 
 /// Call the user-defined mapper function followed by the appropriate
 // target_data_* function (target_data_{begin,end,update}).
 int target_data_mapper(DeviceTy &Device, void *arg_base,
     void *arg, int64_t arg_size, int64_t arg_type, void *arg_mapper,
     TargetDataFuncPtrTy target_data_function) {
int rc = target_data_function(Device, MapperComponents.Components.size(),
       mapper_args_base.data(), mapper_args.data(), mapper_arg_sizes.data(),
       mapper_arg_types.data(), /*arg_mappers*/ nullptr,
       /*__tgt_async_info*/ nullptr);
 
   return rc;
 }
 
 /// Internal function to do the mapping and transfer the data to the device
-int target_data_begin(DeviceTy &Device, int32_t arg_num, void **args_base,
-                      void **args, int64_t *arg_sizes, int64_t *arg_types,
-                      void **arg_mappers, __tgt_async_info *async_info_ptr) {
+int targetDataBegin(DeviceTy &Device, int32_t arg_num, void **args_base,
   // process each input.
   for (int32_t i = 0; i < arg_num; ++i) {
     // Ignore private variables and arrays - there is no mapping for them.
     if ((arg_types[i] & OMP_TGT_MAPTYPE_LITERAL) ||
         (arg_types[i] & OMP_TGT_MAPTYPE_PRIVATE))
       continue;
 
     if (arg_mappers && arg_mappers[i]) {
-      // Instead of executing the regular path of target_data_begin, call the
-      // target_data_mapper variant which will call target_data_begin again
+      // Instead of executing the regular path of targetDataBegin, call the
+      // target_data_mapper variant which will call targetDataBegin again
       // with new arguments.
       DP("Calling target_data_mapper for the %dth argument\n", i);
 
-      int rc = target_data_mapper(Device, args_base[i], args[i], arg_sizes[i],
-          arg_types[i], arg_mappers[i], target_data_begin);
+      int rc =
+          target_data_mapper(Device, args_base[i], args[i], arg_sizes[i],
+                             arg_types[i], arg_mappers[i], targetDataBegin);
 
       if (rc != OFFLOAD_SUCCESS) {
-        DP("Call to target_data_begin via target_data_mapper for custom mapper"
-            " failed.\n");
+        DP("Call to targetDataBegin via target_data_mapper for custom mapper"
         return OFFLOAD_FAIL;
       }
 
       // Skip the rest of this function, continue to the next argument.
       continue;
     }
 
     void *HstPtrBegin = args[i];
     void *HstPtrBase = args_base[i];
     int64_t data_size = arg_sizes[i];
 
     // Adjust for proper alignment if this is a combined entry (for structs).
     // Look at the next argument - if that is MEMBER_OF this one, then this one
     // is a combined entry.
     int64_t padding = 0;
     const int next_i = i+1;
-    if (member_of(arg_types[i]) < 0 && next_i < arg_num &&
-        member_of(arg_types[next_i]) == i) {
+    if (getParentIndex(arg_types[i]) < 0 && next_i < arg_num &&
+        getParentIndex(arg_types[next_i]) == i) {
       padding = (int64_t)HstPtrBegin % alignment;
       if (padding) {
         DP("Using a padding of %" PRId64 " bytes for begin address " DPxMOD
             "\n", padding, DPxPTR(HstPtrBegin));
         HstPtrBegin = (char *) HstPtrBegin - padding;
         data_size += padding;
       }
     }
 
     // Address of pointer on the host and device, respectively.
-    void *Pointer_HstPtrBegin, *Pointer_TgtPtrBegin;
+    void *Pointer_HstPtrBegin, *PointerTgtPtrBegin;
     bool IsNew, Pointer_IsNew;
     bool IsHostPtr = false;
     bool IsImplicit = arg_types[i] & OMP_TGT_MAPTYPE_IMPLICIT;
     // Force the creation of a device side copy of the data when:
     // a close map modifier was associated with a map that contained a to.
     bool HasCloseModifier = arg_types[i] & OMP_TGT_MAPTYPE_CLOSE;
     bool HasPresentModifier = arg_types[i] & OMP_TGT_MAPTYPE_PRESENT;
     // UpdateRef is based on MEMBER_OF instead of TARGET_PARAM because if we
     // have reached this point via __tgt_target_data_begin and not __tgt_target
     // then no argument is marked as TARGET_PARAM ("omp target data map" is not
     // associated with a target region, so there are no target parameters). This
     // may be considered a hack, we could revise the scheme in the future.
     bool UpdateRef = !(arg_types[i] & OMP_TGT_MAPTYPE_MEMBER_OF);
     if (arg_types[i] & OMP_TGT_MAPTYPE_PTR_AND_OBJ) {
       DP("Has a pointer entry: \n");
       // Base is address of pointer.
       //
       // Usually, the pointer is already allocated by this time.  For example:
       //
       //   #pragma omp target map(s.p[0:N])
       //
       // The map entry for s comes first, and the PTR_AND_OBJ entry comes
       // afterward, so the pointer is already allocated by the time the
-      // PTR_AND_OBJ entry is handled below, and Pointer_TgtPtrBegin is thus
+      // PTR_AND_OBJ entry is handled below, and PointerTgtPtrBegin is thus
       // non-null.  However, "declare target link" can produce a PTR_AND_OBJ
       // entry for a global that might not already be allocated by the time the
       // PTR_AND_OBJ entry is handled below, and so the allocation might fail
       // when HasPresentModifier.
-      Pointer_TgtPtrBegin = Device.getOrAllocTgtPtr(
+      PointerTgtPtrBegin = Device.getOrAllocTgtPtr(
           HstPtrBase, HstPtrBase, sizeof(void *), Pointer_IsNew, IsHostPtr,
           IsImplicit, UpdateRef, HasCloseModifier, HasPresentModifier);
-      if (!Pointer_TgtPtrBegin) {
+      if (!PointerTgtPtrBegin) {
         DP("Call to getOrAllocTgtPtr returned null pointer (%s).\n",
            HasPresentModifier ? "'present' map type modifier"
                               : "device failure or illegal mapping");
         return OFFLOAD_FAIL;
       }
       DP("There are %zu bytes allocated at target address " DPxMOD " - is%s new"
-          "\n", sizeof(void *), DPxPTR(Pointer_TgtPtrBegin),
+          "\n", sizeof(void *), DPxPTR(PointerTgtPtrBegin),
           (Pointer_IsNew ? "" : " not"));
       Pointer_HstPtrBegin = HstPtrBase;
       // modify current entry.
       HstPtrBase = *(void **)HstPtrBase;
       UpdateRef = true; // subsequently update ref count of pointee
     }
 
     void *TgtPtrBegin = Device.getOrAllocTgtPtr(
for (int32_t i = 0; i < arg_num; ++i) {
     if (arg_types[i] & OMP_TGT_MAPTYPE_TO) {
       bool copy = false;
       if (!(RTLs->RequiresFlags & OMP_REQ_UNIFIED_SHARED_MEMORY) ||
           HasCloseModifier) {
         if (IsNew || (arg_types[i] & OMP_TGT_MAPTYPE_ALWAYS)) {
           copy = true;
         } else if (arg_types[i] & OMP_TGT_MAPTYPE_MEMBER_OF) {
           // Copy data only if the "parent" struct has RefCount==1.
-          int32_t parent_idx = member_of(arg_types[i]);
+          int32_t parent_idx = getParentIndex(arg_types[i]);
           uint64_t parent_rc = Device.getMapEntryRefCnt(args[parent_idx]);
           assert(parent_rc > 0 && "parent struct not found");
           if (parent_rc == 1) {
             copy = true;
           }
         }
       }
 
       if (copy && !IsHostPtr) {
         DP("Moving %" PRId64 " bytes (hst:" DPxMOD ") -> (tgt:" DPxMOD ")\n",
            data_size, DPxPTR(HstPtrBegin), DPxPTR(TgtPtrBegin));
-        int rt = Device.data_submit(TgtPtrBegin, HstPtrBegin, data_size,
-                                    async_info_ptr);
+        int rt = Device.submitData(TgtPtrBegin, HstPtrBegin, data_size,
         if (rt != OFFLOAD_SUCCESS) {
           DP("Copying data to device failed.\n");
           return OFFLOAD_FAIL;
         }
       }
     }
 
     if (arg_types[i] & OMP_TGT_MAPTYPE_PTR_AND_OBJ && !IsHostPtr) {
       DP("Update pointer (" DPxMOD ") -> [" DPxMOD "]\n",
-          DPxPTR(Pointer_TgtPtrBegin), DPxPTR(TgtPtrBegin));
+         DPxPTR(PointerTgtPtrBegin), DPxPTR(TgtPtrBegin));
       uint64_t Delta = (uint64_t)HstPtrBegin - (uint64_t)HstPtrBase;
       void *TgtPtrBase = (void *)((uint64_t)TgtPtrBegin - Delta);
-      int rt = Device.data_submit(Pointer_TgtPtrBegin, &TgtPtrBase,
-                                  sizeof(void *), async_info_ptr);
+      int rt = Device.submitData(PointerTgtPtrBegin, &TgtPtrBase,
       if (rt != OFFLOAD_SUCCESS) {
         DP("Copying data to device failed.\n");
         return OFFLOAD_FAIL;
       }
       // create shadow pointers for this entry
       Device.ShadowMtx.lock();
-      Device.ShadowPtrMap[Pointer_HstPtrBegin] = {HstPtrBase,
-          Pointer_TgtPtrBegin, TgtPtrBase};
+      Device.ShadowPtrMap[Pointer_HstPtrBegin] = {
+          HstPtrBase, PointerTgtPtrBegin, TgtPtrBase};
       Device.ShadowMtx.unlock();
     }
   }
 
   return OFFLOAD_SUCCESS;
 }
 
 /// Internal function to undo the mapping and retrieve the data from the device.
-int target_data_end(DeviceTy &Device, int32_t arg_num, void **args_base,
-                    void **args, int64_t *arg_sizes, int64_t *arg_types,
-                    void **arg_mappers, __tgt_async_info *async_info_ptr) {
+int targetDataEnd(DeviceTy &Device, int32_t arg_num, void **args_base,
   // process each input.
   for (int32_t i = arg_num - 1; i >= 0; --i) {
     // Ignore private variables and arrays - there is no mapping for them.
     // Also, ignore the use_device_ptr directive, it has no effect here.
     if ((arg_types[i] & OMP_TGT_MAPTYPE_LITERAL) ||
         (arg_types[i] & OMP_TGT_MAPTYPE_PRIVATE))
       continue;
 
     if (arg_mappers && arg_mappers[i]) {
-      // Instead of executing the regular path of target_data_end, call the
-      // target_data_mapper variant which will call target_data_end again
+      // Instead of executing the regular path of targetDataEnd, call the
+      // target_data_mapper variant which will call targetDataEnd again
       // with new arguments.
       DP("Calling target_data_mapper for the %dth argument\n", i);
 
       int rc = target_data_mapper(Device, args_base[i], args[i], arg_sizes[i],
-          arg_types[i], arg_mappers[i], target_data_end);
+                                  arg_types[i], arg_mappers[i], targetDataEnd);
 
       if (rc != OFFLOAD_SUCCESS) {
-        DP("Call to target_data_end via target_data_mapper for custom mapper"
-            " failed.\n");
+        DP("Call to targetDataEnd via target_data_mapper for custom mapper"
         return OFFLOAD_FAIL;
       }
 
       // Skip the rest of this function, continue to the next argument.
       continue;
     }
 
     void *HstPtrBegin = args[i];
     int64_t data_size = arg_sizes[i];
     // Adjust for proper alignment if this is a combined entry (for structs).
     // Look at the next argument - if that is MEMBER_OF this one, then this one
     // is a combined entry.
     int64_t padding = 0;
     const int next_i = i+1;
-    if (member_of(arg_types[i]) < 0 && next_i < arg_num &&
-        member_of(arg_types[next_i]) == i) {
+    if (getParentIndex(arg_types[i]) < 0 && next_i < arg_num &&
+        getParentIndex(arg_types[next_i]) == i) {
       padding = (int64_t)HstPtrBegin % alignment;
       if (padding) {
         DP("Using a padding of %" PRId64 " bytes for begin address " DPxMOD
             "\n", padding, DPxPTR(HstPtrBegin));
         HstPtrBegin = (char *) HstPtrBegin - padding;
         data_size += padding;
       }
     }
if ((arg_types[i] & OMP_TGT_MAPTYPE_FROM) || DelEntry) {
       if (arg_types[i] & OMP_TGT_MAPTYPE_FROM) {
         bool Always = arg_types[i] & OMP_TGT_MAPTYPE_ALWAYS;
         bool CopyMember = false;
         if (!(RTLs->RequiresFlags & OMP_REQ_UNIFIED_SHARED_MEMORY) ||
             HasCloseModifier) {
           if ((arg_types[i] & OMP_TGT_MAPTYPE_MEMBER_OF) &&
               !(arg_types[i] & OMP_TGT_MAPTYPE_PTR_AND_OBJ)) {
             // Copy data only if the "parent" struct has RefCount==1.
-            int32_t parent_idx = member_of(arg_types[i]);
+            int32_t parent_idx = getParentIndex(arg_types[i]);
             uint64_t parent_rc = Device.getMapEntryRefCnt(args[parent_idx]);
             assert(parent_rc > 0 && "parent struct not found");
             if (parent_rc == 1) {
               CopyMember = true;
             }
           }
         }
 
for (int32_t i = arg_num - 1; i >= 0; --i) {
     }
   }
 
   return OFFLOAD_SUCCESS;
 }
 
 /// Internal function to pass data to/from the target.
 // async_info_ptr is currently unused, added here so target_data_update has the
-// same signature as target_data_begin and target_data_end.
+// same signature as targetDataBegin and targetDataEnd.
 int target_data_update(DeviceTy &Device, int32_t arg_num,
     void **args_base, void **args, int64_t *arg_sizes, int64_t *arg_types,
     void **arg_mappers, __tgt_async_info *async_info_ptr) {
   // process each input.
   for (int32_t i = 0; i < arg_num; ++i) {
     if ((arg_types[i] & OMP_TGT_MAPTYPE_LITERAL) ||
         (arg_types[i] & OMP_TGT_MAPTYPE_PRIVATE))
       continue;
if (arg_types[i] & OMP_TGT_MAPTYPE_FROM) {
         *ShadowHstPtrAddr = it->second.HstPtrVal;
       }
       Device.ShadowMtx.unlock();
     }
 
     if (arg_types[i] & OMP_TGT_MAPTYPE_TO) {
       DP("Moving %" PRId64 " bytes (hst:" DPxMOD ") -> (tgt:" DPxMOD ")\n",
           arg_sizes[i], DPxPTR(HstPtrBegin), DPxPTR(TgtPtrBegin));
-      int rt = Device.data_submit(TgtPtrBegin, HstPtrBegin, MapSize, nullptr);
+      int rt = Device.submitData(TgtPtrBegin, HstPtrBegin, MapSize, nullptr);
       if (rt != OFFLOAD_SUCCESS) {
         DP("Copying data to device failed.\n");
         return OFFLOAD_FAIL;
       }
 
       uintptr_t lb = (uintptr_t) HstPtrBegin;
       uintptr_t ub = (uintptr_t) HstPtrBegin + MapSize;
       Device.ShadowMtx.lock();
       for (ShadowPtrListTy::iterator it = Device.ShadowPtrMap.begin();
           it != Device.ShadowPtrMap.end(); ++it) {
-        void **ShadowHstPtrAddr = (void**) it->first;
-        if ((uintptr_t) ShadowHstPtrAddr < lb)
+        void **ShadowHstPtrAddr = (void **)it->first;
+        if ((uintptr_t)ShadowHstPtrAddr < lb)
           continue;
-        if ((uintptr_t) ShadowHstPtrAddr >= ub)
+        if ((uintptr_t)ShadowHstPtrAddr >= ub)
           break;
         DP("Restoring original target pointer value " DPxMOD " for target "
-            "pointer " DPxMOD "\n", DPxPTR(it->second.TgtPtrVal),
-            DPxPTR(it->second.TgtPtrAddr));
-        rt = Device.data_submit(it->second.TgtPtrAddr,
-            &it->second.TgtPtrVal, sizeof(void *), nullptr);
+           "pointer " DPxMOD "\n",
+           DPxPTR(it->second.TgtPtrVal), DPxPTR(it->second.TgtPtrAddr));
+        rt = Device.submitData(it->second.TgtPtrAddr, &it->second.TgtPtrVal,
+                               sizeof(void *), nullptr);
         if (rt != OFFLOAD_SUCCESS) {
           DP("Copying data to device failed.\n");
           Device.ShadowMtx.unlock();
           return OFFLOAD_FAIL;
         }
       }
       Device.ShadowMtx.unlock();
     }

 }
 
 /// performs the same actions as data_begin in case arg_num is
 /// non-zero and initiates run of the offloaded region on the target platform;
 /// if arg_num is non-zero after the region execution is done it also
 /// performs the same action as data_update and data_end above. This function
 /// returns 0 if it was able to transfer the execution to a target and an
 /// integer different from zero otherwise.
-int target(int64_t device_id, void *host_ptr, int32_t arg_num,
-    void **args_base, void **args, int64_t *arg_sizes, int64_t *arg_types,
-    void **arg_mappers, int32_t team_num, int32_t thread_limit,
-    int IsTeamConstruct) {
-  DeviceTy &Device = Devices[device_id];
+int target(int64_t DeviceId, void *HostPtr, int32_t ArgNum, void **ArgBases,
+           void **Args, int64_t *ArgSizes, int64_t *ArgTypes, void **ArgMappers,
+           int32_t TeamNum, int32_t ThreadLimit, int IsTeamConstruct) {
+  DeviceTy &Device = Devices[DeviceId];
 
   // Find the table information in the map or look it up in the translation
   // tables.
   TableMap *TM = 0;
   TblMapMtx->lock();
-  HostPtrToTableMapTy::iterator TableMapIt = HostPtrToTableMap->find(host_ptr);
+  HostPtrToTableMapTy::iterator TableMapIt = HostPtrToTableMap->find(HostPtr);
   if (TableMapIt == HostPtrToTableMap->end()) {
     // We don't have a map. So search all the registered libraries.
     TrlTblMtx->lock();
     for (HostEntriesBeginToTransTableTy::iterator
-             ii = HostEntriesBeginToTransTable->begin(),
-             ie = HostEntriesBeginToTransTable->end();
-         !TM && ii != ie; ++ii) {
+             II = HostEntriesBeginToTransTable->begin(),
+             IE = HostEntriesBeginToTransTable->end();
+         !TM && II != IE; ++II) {
       // get the translation table (which contains all the good info).
-      TranslationTable *TransTable = &ii->second;
+      TranslationTable *TransTable = &II->second;
       // iterate over all the host table entries to see if we can locate the
       // host_ptr.
-      __tgt_offload_entry *begin = TransTable->HostTable.EntriesBegin;
-      __tgt_offload_entry *end = TransTable->HostTable.EntriesEnd;
-      __tgt_offload_entry *cur = begin;
-      for (uint32_t i = 0; cur < end; ++cur, ++i) {
-        if (cur->addr != host_ptr)
+      __tgt_offload_entry *Begin = TransTable->HostTable.EntriesBegin;
+      __tgt_offload_entry *End = TransTable->HostTable.EntriesEnd;
+      __tgt_offload_entry *Cur = Begin;
+      for (uint32_t I = 0; Cur < End; ++Cur, ++I) {
+        if (Cur->addr != HostPtr)
           continue;
         // we got a match, now fill the HostPtrToTableMap so that we
         // may avoid this search next time.
-        TM = &(*HostPtrToTableMap)[host_ptr];
+        TM = &(*HostPtrToTableMap)[HostPtr];
         TM->Table = TransTable;
-        TM->Index = i;
+        TM->Index = I;
         break;
       }
     }
     TrlTblMtx->unlock();
   } else {
     TM = &TableMapIt->second;
   }
   TblMapMtx->unlock();
 
   // No map for this host pointer found!
   if (!TM) {
     DP("Host ptr " DPxMOD " does not have a matching target pointer.\n",
-       DPxPTR(host_ptr));
+       DPxPTR(HostPtr));
     return OFFLOAD_FAIL;
   }
 
   // get target table.
   TrlTblMtx->lock();
-  assert(TM->Table->TargetsTable.size() > (size_t)device_id &&
+  assert(TM->Table->TargetsTable.size() > (size_t)DeviceId &&
          "Not expecting a device ID outside the table's bounds!");
-  __tgt_target_table *TargetTable = TM->Table->TargetsTable[device_id];
+  __tgt_target_table *TargetTable = TM->Table->TargetsTable[DeviceId];
   TrlTblMtx->unlock();
   assert(TargetTable && "Global data has not been mapped\n");
 
   __tgt_async_info AsyncInfo;
 
   // Move data to device.
-  int rc = target_data_begin(Device, arg_num, args_base, args, arg_sizes,
-                             arg_types, arg_mappers, &AsyncInfo);
-  if (rc != OFFLOAD_SUCCESS) {
-    DP("Call to target_data_begin failed, abort target.\n");
+  int Ret = targetDataBegin(Device, ArgNum, ArgBases, Args, ArgSizes, ArgTypes,
+                            ArgMappers, &AsyncInfo);
+  if (Ret != OFFLOAD_SUCCESS) {
+    DP("Call to targetDataBegin failed, abort target.\n");
     return OFFLOAD_FAIL;
   }
 
-  std::vector<void *> tgt_args;
-  std::vector<ptrdiff_t> tgt_offsets;
+  std::vector<void *> TgtArgs;
+  std::vector<ptrdiff_t> TgtOffsets;
 
   // List of (first-)private arrays allocated for this target region
-  std::vector<void *> fpArrays;
-  std::vector<int> tgtArgsPositions(arg_num, -1);
+  std::vector<void *> FPArrays;
+  std::vector<int> TgtArgsPositions(ArgNum, -1);
 
-  for (int32_t i = 0; i < arg_num; ++i) {
-    if (!(arg_types[i] & OMP_TGT_MAPTYPE_TARGET_PARAM)) {
-      // This is not a target parameter, do not push it into tgt_args.
+  for (int32_t I = 0; I < ArgNum; ++I) {
+    if (!(ArgTypes[I] & OMP_TGT_MAPTYPE_TARGET_PARAM)) {
+      // This is not a target parameter, do not push it into TgtArgs.
       // Check for lambda mapping.
-      if (isLambdaMapping(arg_types[i])) {
-        assert((arg_types[i] & OMP_TGT_MAPTYPE_MEMBER_OF) &&
+      if (isLambdaMapping(ArgTypes[I])) {
+        assert((ArgTypes[I] & OMP_TGT_MAPTYPE_MEMBER_OF) &&
                "PTR_AND_OBJ must be also MEMBER_OF.");
-        unsigned idx = member_of(arg_types[i]);
-        int tgtIdx = tgtArgsPositions[idx];
-        assert(tgtIdx != -1 && "Base address must be translated already.");
+        unsigned Idx = getParentIndex(ArgTypes[I]);
+        int TgtIdx = TgtArgsPositions[Idx];
+        assert(TgtIdx != -1 && "Base address must be translated already.");
         // The parent lambda must be processed already and it must be the last
-        // in tgt_args and tgt_offsets arrays.
-        void *HstPtrVal = args[i];
-        void *HstPtrBegin = args_base[i];
-        void *HstPtrBase = args[idx];
+        // in TgtArgs and TgtOffsets arrays.
+        void *HstPtrVal = Args[I];
+        void *HstPtrBegin = ArgBases[I];
+        void *HstPtrBase = Args[Idx];
         bool IsLast, IsHostPtr; // unused.
         void *TgtPtrBase =
-            (void *)((intptr_t)tgt_args[tgtIdx] + tgt_offsets[tgtIdx]);
+            (void *)((intptr_t)TgtArgs[TgtIdx] + TgtOffsets[TgtIdx]);
         DP("Parent lambda base " DPxMOD "\n", DPxPTR(TgtPtrBase));
         uint64_t Delta = (uint64_t)HstPtrBegin - (uint64_t)HstPtrBase;
         void *TgtPtrBegin = (void *)((uintptr_t)TgtPtrBase + Delta);
-        void *Pointer_TgtPtrBegin =
-            Device.getTgtPtrBegin(HstPtrVal, arg_sizes[i], IsLast, false,
-                                  IsHostPtr);
-        if (!Pointer_TgtPtrBegin) {
+        void *PointerTgtPtrBegin = Device.getTgtPtrBegin(
+            HstPtrVal, ArgSizes[I], IsLast, false, IsHostPtr);
+        if (!PointerTgtPtrBegin) {
           DP("No lambda captured variable mapped (" DPxMOD ") - ignored\n",
              DPxPTR(HstPtrVal));
           continue;
         }
         if (RTLs->RequiresFlags & OMP_REQ_UNIFIED_SHARED_MEMORY &&
             TgtPtrBegin == HstPtrBegin) {
           DP("Unified memory is active, no need to map lambda captured"
-             "variable (" DPxMOD ")\n", DPxPTR(HstPtrVal));
+             "variable (" DPxMOD ")\n",
+             DPxPTR(HstPtrVal));
           continue;
         }
         DP("Update lambda reference (" DPxMOD ") -> [" DPxMOD "]\n",
-           DPxPTR(Pointer_TgtPtrBegin), DPxPTR(TgtPtrBegin));
-        int rt = Device.data_submit(TgtPtrBegin, &Pointer_TgtPtrBegin,
-                                    sizeof(void *), &AsyncInfo);
-        if (rt != OFFLOAD_SUCCESS) {
+           DPxPTR(PointerTgtPtrBegin), DPxPTR(TgtPtrBegin));
+        Ret = Device.submitData(TgtPtrBegin, &PointerTgtPtrBegin,
+        if (Ret != OFFLOAD_SUCCESS) {
           DP("Copying data to device failed.\n");
           return OFFLOAD_FAIL;
         }
       }
       continue;
     }
-    void *HstPtrBegin = args[i];
-    void *HstPtrBase = args_base[i];
+    void *HstPtrBegin = Args[I];
+    void *HstPtrBase = ArgBases[I];
     void *TgtPtrBegin;
     ptrdiff_t TgtBaseOffset;
     bool IsLast, IsHostPtr; // unused.
-    if (arg_types[i] & OMP_TGT_MAPTYPE_LITERAL) {
+    if (ArgTypes[I] & OMP_TGT_MAPTYPE_LITERAL) {
       DP("Forwarding first-private value " DPxMOD " to the target construct\n",
-          DPxPTR(HstPtrBase));
       TgtPtrBegin = HstPtrBase;
       TgtBaseOffset = 0;
-    } else if (arg_types[i] & OMP_TGT_MAPTYPE_PRIVATE) {
+    } else if (ArgTypes[I] & OMP_TGT_MAPTYPE_PRIVATE) {
       // Allocate memory for (first-)private array
-      TgtPtrBegin = Device.data_alloc(arg_sizes[i], HstPtrBegin);
+      TgtPtrBegin = Device.allocData(ArgSizes[I], HstPtrBegin);
       if (!TgtPtrBegin) {
-        DP ("Data allocation for %sprivate array " DPxMOD " failed, "
-            "abort target.\n",
-            (arg_types[i] & OMP_TGT_MAPTYPE_TO ? "first-" : ""),
-            DPxPTR(HstPtrBegin));
+        DP("Data allocation for %sprivate array " DPxMOD " failed, "
+           (ArgTypes[I] & OMP_TGT_MAPTYPE_TO ? "first-" : ""),
         return OFFLOAD_FAIL;
       }
-      fpArrays.push_back(TgtPtrBegin);
+      FPArrays.push_back(TgtPtrBegin);
       TgtBaseOffset = (intptr_t)HstPtrBase - (intptr_t)HstPtrBegin;
 #ifdef OMPTARGET_DEBUG
       void *TgtPtrBase = (void *)((intptr_t)TgtPtrBegin + TgtBaseOffset);
       DP("Allocated %" PRId64 " bytes of target memory at " DPxMOD " for "
-          "%sprivate array " DPxMOD " - pushing target argument " DPxMOD "\n",
-          arg_sizes[i], DPxPTR(TgtPtrBegin),
-          (arg_types[i] & OMP_TGT_MAPTYPE_TO ? "first-" : ""),
-          DPxPTR(HstPtrBegin), DPxPTR(TgtPtrBase));
+         ArgSizes[I], DPxPTR(TgtPtrBegin),
+         (ArgTypes[I] & OMP_TGT_MAPTYPE_TO ? "first-" : ""),
 #endif
       // If first-private, copy data from host
-      if (arg_types[i] & OMP_TGT_MAPTYPE_TO) {
-        int rt = Device.data_submit(TgtPtrBegin, HstPtrBegin, arg_sizes[i],
-                                    &AsyncInfo);
-        if (rt != OFFLOAD_SUCCESS) {
+      if (ArgTypes[I] & OMP_TGT_MAPTYPE_TO) {
+        Ret = Device.submitData(TgtPtrBegin, HstPtrBegin, ArgSizes[I],
+        if (Ret != OFFLOAD_SUCCESS) {
           DP("Copying data to device failed, failed.\n");
           return OFFLOAD_FAIL;
         }
       }
-    } else if (arg_types[i] & OMP_TGT_MAPTYPE_PTR_AND_OBJ) {
+    } else if (ArgTypes[I] & OMP_TGT_MAPTYPE_PTR_AND_OBJ) {
       TgtPtrBegin = Device.getTgtPtrBegin(HstPtrBase, sizeof(void *), IsLast,
                                           false, IsHostPtr);
       TgtBaseOffset = 0; // no offset for ptrs.
       DP("Obtained target argument " DPxMOD " from host pointer " DPxMOD " to "
-         "object " DPxMOD "\n", DPxPTR(TgtPtrBegin), DPxPTR(HstPtrBase),
-         DPxPTR(HstPtrBase));
+         "object " DPxMOD "\n",
+         DPxPTR(TgtPtrBegin), DPxPTR(HstPtrBase), DPxPTR(HstPtrBase));
     } else {
-      TgtPtrBegin = Device.getTgtPtrBegin(HstPtrBegin, arg_sizes[i], IsLast,
+      TgtPtrBegin = Device.getTgtPtrBegin(HstPtrBegin, ArgSizes[I], IsLast,
                                           false, IsHostPtr);
       TgtBaseOffset = (intptr_t)HstPtrBase - (intptr_t)HstPtrBegin;
 #ifdef OMPTARGET_DEBUG
       void *TgtPtrBase = (void *)((intptr_t)TgtPtrBegin + TgtBaseOffset);
       DP("Obtained target argument " DPxMOD " from host pointer " DPxMOD "\n",
-          DPxPTR(TgtPtrBase), DPxPTR(HstPtrBegin));
 #endif
     }
-    tgtArgsPositions[i] = tgt_args.size();
-    tgt_args.push_back(TgtPtrBegin);
-    tgt_offsets.push_back(TgtBaseOffset);
+    TgtArgsPositions[I] = TgtArgs.size();
+    TgtArgs.push_back(TgtPtrBegin);
+    TgtOffsets.push_back(TgtBaseOffset);
   }
 
-  assert(tgt_args.size() == tgt_offsets.size() &&
+  assert(TgtArgs.size() == TgtOffsets.size() &&
          "Size mismatch in arguments and offsets");
 
   // Pop loop trip count
-  uint64_t ltc = 0;
+  uint64_t LoopTripCount = 0;
   TblMapMtx->lock();
   auto I = Device.LoopTripCnt.find(__kmpc_global_thread_num(NULL));
   if (I != Device.LoopTripCnt.end()) {
-    ltc = I->second;
+    LoopTripCount = I->second;
     Device.LoopTripCnt.erase(I);
-    DP("loop trip count is %lu.\n", ltc);
+    DP("loop trip count is %lu.\n", LoopTripCount);
   }
   TblMapMtx->unlock();
 
   // Launch device execution.
   DP("Launching target execution %s with pointer " DPxMOD " (index=%d).\n",
-      TargetTable->EntriesBegin[TM->Index].name,
-      DPxPTR(TargetTable->EntriesBegin[TM->Index].addr), TM->Index);
   if (IsTeamConstruct) {
-    rc = Device.run_team_region(TargetTable->EntriesBegin[TM->Index].addr,
-                                &tgt_args[0], &tgt_offsets[0], tgt_args.size(),
-                                team_num, thread_limit, ltc, &AsyncInfo);
+    Ret = Device.runTeamRegion(TargetTable->EntriesBegin[TM->Index].addr,
+                               &TgtArgs[0], &TgtOffsets[0], TgtArgs.size(),
+                               TeamNum, ThreadLimit, LoopTripCount, &AsyncInfo);
   } else {
-    rc = Device.run_region(TargetTable->EntriesBegin[TM->Index].addr,
-                           &tgt_args[0], &tgt_offsets[0], tgt_args.size(),
-                           &AsyncInfo);
+    Ret =
+        Device.runRegion(TargetTable->EntriesBegin[TM->Index].addr, &TgtArgs[0],
+                         &TgtOffsets[0], TgtArgs.size(), &AsyncInfo);
   }
-  if (rc != OFFLOAD_SUCCESS) {
-    DP ("Executing target region abort target.\n");
+  if (Ret != OFFLOAD_SUCCESS) {
+    DP("Executing target region abort target.\n");
     return OFFLOAD_FAIL;
   }
 
   // Deallocate (first-)private arrays
-  for (auto it : fpArrays) {
-    int rt = Device.data_delete(it);
-    if (rt != OFFLOAD_SUCCESS) {
+  for (auto Itr : FPArrays) {
+    Ret = Device.deleteData(Itr);
+    if (Ret != OFFLOAD_SUCCESS) {
       DP("Deallocation of (first-)private arrays failed.\n");
       return OFFLOAD_FAIL;
     }
   }
 
   // Move data from device.
-  int rt = target_data_end(Device, arg_num, args_base, args, arg_sizes,
-                           arg_types, arg_mappers, &AsyncInfo);
-  if (rt != OFFLOAD_SUCCESS) {
-    DP("Call to target_data_end failed, abort targe.\n");
+  Ret = targetDataEnd(Device, ArgNum, ArgBases, Args, ArgSizes, ArgTypes,
+                      ArgMappers, &AsyncInfo);
+  if (Ret != OFFLOAD_SUCCESS) {
+    DP("Call to targetDataEnd failed, abort targe.\n");
     return OFFLOAD_FAIL;
   }
 
   return Device.synchronize(&AsyncInfo);
 }

openmp/libomptarget/src/private.h

 
 #ifndef _OMPTARGET_PRIVATE_H
 #define _OMPTARGET_PRIVATE_H
 
 #include <omptarget.h>
 
 #include <cstdint>
 
-extern int target_data_begin(DeviceTy &Device, int32_t arg_num,
-                             void **args_base, void **args, int64_t *arg_sizes,
-                             int64_t *arg_types, void **arg_mappers,
-                             __tgt_async_info *async_info_ptr);
-
-extern int target_data_end(DeviceTy &Device, int32_t arg_num, void **args_base,
+extern int targetDataBegin(DeviceTy &Device, int32_t arg_num, void **args_base,
                            void **args, int64_t *arg_sizes, int64_t *arg_types,
                            void **arg_mappers,
                            __tgt_async_info *async_info_ptr);
 
+extern int targetDataEnd(DeviceTy &Device, int32_t arg_num, void **args_base,
+                         void **args, int64_t *arg_sizes, int64_t *arg_types,
+                         void **arg_mappers, __tgt_async_info *async_info_ptr);
+
 extern int target_data_update(DeviceTy &Device, int32_t arg_num,
                               void **args_base, void **args,
                               int64_t *arg_sizes, int64_t *arg_types,
                               void **arg_mappers,
                               __tgt_async_info *async_info_ptr = nullptr);
 
-extern int target(int64_t device_id, void *host_ptr, int32_t arg_num,
-                  void **args_base, void **args, int64_t *arg_sizes,
-                  int64_t *arg_types, void **arg_mappers, int32_t team_num,
-                  int32_t thread_limit, int IsTeamConstruct);
+extern int target(int64_t DeviceId, void *HostPtr, int32_t ArgNum,
+                  void **ArgBases, void **Args, int64_t *ArgSizes,
+                  int64_t *ArgTypes, void **ArgMappers, int32_t TeamNum,
+                  int32_t ThreadLimit, int IsTeamConstruct);
 
 extern int CheckDeviceAndCtors(int64_t device_id);
 
 // enum for OMP_TARGET_OFFLOAD; keep in sync with kmp.h definition
 enum kmp_target_offload_kind {
   tgt_disabled = 0,
   tgt_default = 1,
   tgt_mandatory = 2

 };
 
 // The mapper function pointer type. It follows the signature below:
 // void .omp_mapper.<type_name>.<mapper_id>.(void *rt_mapper_handle,
 //                                           void *base, void *begin,
 //                                           size_t size, int64_t type);
 typedef void (*MapperFuncPtrTy)(void *, void *, void *, int64_t, int64_t);
 
-// Function pointer type for target_data_* functions (target_data_begin,
-// target_data_end and target_data_update).
+// Function pointer type for target_data_* functions (targetDataBegin,
+// targetDataEnd and target_data_update).
 typedef int (*TargetDataFuncPtrTy)(DeviceTy &, int32_t, void **, void **,
-    int64_t *, int64_t *, void **, __tgt_async_info *);
+                                   int64_t *, int64_t *, void **,
+                                   __tgt_async_info *);
 
 ////////////////////////////////////////////////////////////////////////////////
 // implementation for messages
 ////////////////////////////////////////////////////////////////////////////////
 
 #define MESSAGE0(_str)                                                         \
   do {                                                                         \
     fprintf(stderr, "Libomptarget message: %s\n", _str);                       \