[Openmp-commits] [openmp] [Libomptarget][NFC] Format in-line comments consistently (PR #77530)

Joseph Huber via Openmp-commits openmp-commits at lists.llvm.org
Wed Jan 10 08:10:11 PST 2024


https://github.com/jhuber6 updated https://github.com/llvm/llvm-project/pull/77530

>From 452744131e8d353fa556cd2aebf11f425b564aef Mon Sep 17 00:00:00 2001
From: Joseph Huber <huberjn at outlook.com>
Date: Tue, 9 Jan 2024 15:54:48 -0600
Subject: [PATCH] [Libomptarget][NFC] Format in-line comments consistently

Summary:
The LLVM style uses /*Foo=*/ when indicating the name of a constant. See
https://llvm.org/docs/CodingStandards.html#comment-formatting. This is
useful for consistency, as well as because `clang-format` understands
this syntax and formats it more cleanly. Do a bulk update of this
syntax.
---
 openmp/libomptarget/DeviceRTL/include/State.h | 26 +++++-----
 openmp/libomptarget/DeviceRTL/src/Kernel.cpp  |  4 +-
 .../DeviceRTL/src/Parallelism.cpp             | 26 +++++-----
 .../libomptarget/DeviceRTL/src/Reduction.cpp  |  2 +-
 openmp/libomptarget/include/Shared/Profile.h  |  2 +-
 .../plugins-nextgen/amdgpu/src/rtl.cpp        | 14 +++---
 .../common/include/GlobalHandler.h            |  8 ++--
 .../plugins-nextgen/common/src/JIT.cpp        |  8 ++--
 .../common/src/PluginInterface.cpp            | 48 ++++++++-----------
 .../plugins-nextgen/cuda/src/rtl.cpp          |  8 ++--
 .../generic-elf-64bit/src/rtl.cpp             |  4 +-
 openmp/libomptarget/src/OpenMP/Mapping.cpp    | 18 +++----
 .../libomptarget/src/OpenMP/OMPT/Callback.cpp | 36 +++++++-------
 openmp/libomptarget/src/device.cpp            |  8 ++--
 openmp/libomptarget/src/interface.cpp         |  6 +--
 openmp/libomptarget/src/omptarget.cpp         |  6 +--
 .../kernelreplay/llvm-omp-kernel-replay.cpp   | 24 +++++-----
 17 files changed, 120 insertions(+), 128 deletions(-)

diff --git a/openmp/libomptarget/DeviceRTL/include/State.h b/openmp/libomptarget/DeviceRTL/include/State.h
index c93de4191f83fa..1a3490394458ff 100644
--- a/openmp/libomptarget/DeviceRTL/include/State.h
+++ b/openmp/libomptarget/DeviceRTL/include/State.h
@@ -240,29 +240,29 @@ lookupPtr(ValueKind Kind, bool IsReadonly, bool ForceTeamState) {
 /// update ICV values we can declare in global scope.
 template <typename Ty, ValueKind Kind> struct Value {
   [[gnu::flatten, gnu::always_inline]] operator Ty() {
-    return lookup(/* IsReadonly */ true, /* IdentTy */ nullptr,
-                  /* ForceTeamState */ false);
+    return lookup(/*IsReadonly=*/true, /*IdentTy=*/nullptr,
+                  /*ForceTeamState=*/false);
   }
 
   [[gnu::flatten, gnu::always_inline]] Value &operator=(const Ty &Other) {
-    set(Other, /* IdentTy */ nullptr);
+    set(Other, /*IdentTy=*/nullptr);
     return *this;
   }
 
   [[gnu::flatten, gnu::always_inline]] Value &operator++() {
-    inc(1, /* IdentTy */ nullptr);
+    inc(1, /*IdentTy=*/nullptr);
     return *this;
   }
 
   [[gnu::flatten, gnu::always_inline]] Value &operator--() {
-    inc(-1, /* IdentTy */ nullptr);
+    inc(-1, /*IdentTy=*/nullptr);
     return *this;
   }
 
   [[gnu::flatten, gnu::always_inline]] void
   assert_eq(const Ty &V, IdentTy *Ident = nullptr,
             bool ForceTeamState = false) {
-    ASSERT(lookup(/* IsReadonly */ true, Ident, ForceTeamState) == V, nullptr);
+    ASSERT(lookup(/*IsReadonly=*/true, Ident, ForceTeamState) == V, nullptr);
   }
 
 private:
@@ -273,12 +273,12 @@ template <typename Ty, ValueKind Kind> struct Value {
   }
 
   [[gnu::flatten, gnu::always_inline]] Ty &inc(int UpdateVal, IdentTy *Ident) {
-    return (lookup(/* IsReadonly */ false, Ident, /* ForceTeamState */ false) +=
+    return (lookup(/*IsReadonly=*/false, Ident, /*ForceTeamState=*/false) +=
             UpdateVal);
   }
 
   [[gnu::flatten, gnu::always_inline]] Ty &set(Ty UpdateVal, IdentTy *Ident) {
-    return (lookup(/* IsReadonly */ false, Ident, /* ForceTeamState */ false) =
+    return (lookup(/*IsReadonly=*/false, Ident, /*ForceTeamState=*/false) =
                 UpdateVal);
   }
 
@@ -290,8 +290,8 @@ template <typename Ty, ValueKind Kind> struct Value {
 /// we can declare in global scope.
 template <typename Ty, ValueKind Kind> struct PtrValue {
   [[gnu::flatten, gnu::always_inline]] operator Ty() {
-    return lookup(/* IsReadonly */ true, /* IdentTy */ nullptr,
-                  /* ForceTeamState */ false);
+    return lookup(/*IsReadonly=*/true, /*IdentTy=*/nullptr,
+                  /*ForceTeamState=*/false);
   }
 
   [[gnu::flatten, gnu::always_inline]] PtrValue &operator=(const Ty Other) {
@@ -305,8 +305,8 @@ template <typename Ty, ValueKind Kind> struct PtrValue {
   }
 
   Ty &set(Ty UpdateVal) {
-    return (lookup(/* IsReadonly */ false, /* IdentTy */ nullptr,
-                   /* ForceTeamState */ false) = UpdateVal);
+    return (lookup(/*IsReadonly=*/false, /*IdentTy=*/nullptr,
+                   /*ForceTeamState=*/false) = UpdateVal);
   }
 
   template <typename VTy, typename Ty2> friend struct ValueRAII;
@@ -315,7 +315,7 @@ template <typename Ty, ValueKind Kind> struct PtrValue {
 template <typename VTy, typename Ty> struct ValueRAII {
   ValueRAII(VTy &V, Ty NewValue, Ty OldValue, bool Active, IdentTy *Ident,
             bool ForceTeamState = false)
-      : Ptr(Active ? &V.lookup(/* IsReadonly */ false, Ident, ForceTeamState)
+      : Ptr(Active ? &V.lookup(/*IsReadonly=*/false, Ident, ForceTeamState)
                    : (Ty *)utils::UndefPtr),
         Val(OldValue), Active(Active) {
     if (!Active)
diff --git a/openmp/libomptarget/DeviceRTL/src/Kernel.cpp b/openmp/libomptarget/DeviceRTL/src/Kernel.cpp
index 06b12fec21677c..95d4c728016d24 100644
--- a/openmp/libomptarget/DeviceRTL/src/Kernel.cpp
+++ b/openmp/libomptarget/DeviceRTL/src/Kernel.cpp
@@ -78,11 +78,11 @@ int32_t __kmpc_target_init(KernelEnvironmentTy &KernelEnvironment,
                 llvm::omp::OMPTgtExecModeFlags::OMP_TGT_EXEC_MODE_SPMD;
   bool UseGenericStateMachine = Configuration.UseGenericStateMachine;
   if (IsSPMD) {
-    inititializeRuntime(/* IsSPMD */ true, KernelEnvironment,
+    inititializeRuntime(/*IsSPMD=*/true, KernelEnvironment,
                         KernelLaunchEnvironment);
     synchronize::threadsAligned(atomic::relaxed);
   } else {
-    inititializeRuntime(/* IsSPMD */ false, KernelEnvironment,
+    inititializeRuntime(/*IsSPMD=*/false, KernelEnvironment,
                         KernelLaunchEnvironment);
     // No need to wait since only the main threads will execute user
     // code and workers will run into a barrier right away.
diff --git a/openmp/libomptarget/DeviceRTL/src/Parallelism.cpp b/openmp/libomptarget/DeviceRTL/src/Parallelism.cpp
index 2c0701bd5358fd..7005477bf4c79d 100644
--- a/openmp/libomptarget/DeviceRTL/src/Parallelism.cpp
+++ b/openmp/libomptarget/DeviceRTL/src/Parallelism.cpp
@@ -121,20 +121,20 @@ __kmpc_parallel_51(IdentTy *ident, int32_t, int32_t if_expr,
       // created.
       state::ValueRAII ParallelTeamSizeRAII(state::ParallelTeamSize, PTeamSize,
                                             1u, TId == 0, ident,
-                                            /* ForceTeamState */ true);
+                                            /*ForceTeamState=*/true);
       state::ValueRAII ActiveLevelRAII(icv::ActiveLevel, 1u, 0u, TId == 0,
-                                       ident, /* ForceTeamState */ true);
+                                       ident, /*ForceTeamState=*/true);
       state::ValueRAII LevelRAII(icv::Level, 1u, 0u, TId == 0, ident,
-                                 /* ForceTeamState */ true);
+                                 /*ForceTeamState=*/true);
 
       // Synchronize all threads after the main thread (TId == 0) set up the
       // team state properly.
       synchronize::threadsAligned(atomic::acq_rel);
 
       state::ParallelTeamSize.assert_eq(PTeamSize, ident,
-                                        /* ForceTeamState */ true);
-      icv::ActiveLevel.assert_eq(1u, ident, /* ForceTeamState */ true);
-      icv::Level.assert_eq(1u, ident, /* ForceTeamState */ true);
+                                        /*ForceTeamState=*/true);
+      icv::ActiveLevel.assert_eq(1u, ident, /*ForceTeamState=*/true);
+      icv::Level.assert_eq(1u, ident, /*ForceTeamState=*/true);
 
       // Ensure we synchronize before we run user code to avoid invalidating the
       // assumptions above.
@@ -152,9 +152,9 @@ __kmpc_parallel_51(IdentTy *ident, int32_t, int32_t if_expr,
     // __kmpc_target_deinit may not hold.
     synchronize::threadsAligned(atomic::acq_rel);
 
-    state::ParallelTeamSize.assert_eq(1u, ident, /* ForceTeamState */ true);
-    icv::ActiveLevel.assert_eq(0u, ident, /* ForceTeamState */ true);
-    icv::Level.assert_eq(0u, ident, /* ForceTeamState */ true);
+    state::ParallelTeamSize.assert_eq(1u, ident, /*ForceTeamState=*/true);
+    icv::ActiveLevel.assert_eq(0u, ident, /*ForceTeamState=*/true);
+    icv::Level.assert_eq(0u, ident, /*ForceTeamState=*/true);
 
     // Ensure we synchronize to create an aligned region around the assumptions.
     synchronize::threadsAligned(atomic::relaxed);
@@ -242,14 +242,14 @@ __kmpc_parallel_51(IdentTy *ident, int32_t, int32_t if_expr,
     // created.
     state::ValueRAII ParallelTeamSizeRAII(state::ParallelTeamSize, PTeamSize,
                                           1u, true, ident,
-                                          /* ForceTeamState */ true);
+                                          /*ForceTeamState=*/true);
     state::ValueRAII ParallelRegionFnRAII(state::ParallelRegionFn, wrapper_fn,
                                           (void *)nullptr, true, ident,
-                                          /* ForceTeamState */ true);
+                                          /*ForceTeamState=*/true);
     state::ValueRAII ActiveLevelRAII(icv::ActiveLevel, 1u, 0u, true, ident,
-                                     /* ForceTeamState */ true);
+                                     /*ForceTeamState=*/true);
     state::ValueRAII LevelRAII(icv::Level, 1u, 0u, true, ident,
-                               /* ForceTeamState */ true);
+                               /*ForceTeamState=*/true);
 
     // Master signals work to activate workers.
     synchronize::threads(atomic::seq_cst);
diff --git a/openmp/libomptarget/DeviceRTL/src/Reduction.cpp b/openmp/libomptarget/DeviceRTL/src/Reduction.cpp
index 0a9db3edabccbf..744d1a3a231c83 100644
--- a/openmp/libomptarget/DeviceRTL/src/Reduction.cpp
+++ b/openmp/libomptarget/DeviceRTL/src/Reduction.cpp
@@ -69,7 +69,7 @@ static int32_t nvptx_parallel_reduce_nowait(void *reduce_data,
                                             ShuffleReductFnTy shflFct,
                                             InterWarpCopyFnTy cpyFct) {
   uint32_t BlockThreadId = mapping::getThreadIdInBlock();
-  if (mapping::isMainThreadInGenericMode(/* IsSPMD */ false))
+  if (mapping::isMainThreadInGenericMode(/*IsSPMD=*/false))
     BlockThreadId = 0;
   uint32_t NumThreads = omp_get_num_threads();
   if (NumThreads == 1)
diff --git a/openmp/libomptarget/include/Shared/Profile.h b/openmp/libomptarget/include/Shared/Profile.h
index 7e580988a39baf..39817bab36cbb2 100644
--- a/openmp/libomptarget/include/Shared/Profile.h
+++ b/openmp/libomptarget/include/Shared/Profile.h
@@ -32,7 +32,7 @@ class Profiler {
     Int32Envar ProfileGranularity =
         Int32Envar("LIBOMPTARGET_PROFILE_GRANULARITY", 500);
 
-    llvm::timeTraceProfilerInitialize(ProfileGranularity /* us */,
+    llvm::timeTraceProfilerInitialize(ProfileGranularity /*us=*/,
                                       "libomptarget");
   }
 
diff --git a/openmp/libomptarget/plugins-nextgen/amdgpu/src/rtl.cpp b/openmp/libomptarget/plugins-nextgen/amdgpu/src/rtl.cpp
index b67642e9e1bcb3..8424d0f5df0822 100644
--- a/openmp/libomptarget/plugins-nextgen/amdgpu/src/rtl.cpp
+++ b/openmp/libomptarget/plugins-nextgen/amdgpu/src/rtl.cpp
@@ -2211,10 +2211,9 @@ struct AMDGPUDeviceTy : public GenericDeviceTy, AMDGenericDeviceTy {
     hsa_amd_pointer_info_t Info;
     Info.size = sizeof(hsa_amd_pointer_info_t);
 
-    hsa_status_t Status =
-        hsa_amd_pointer_info(HstPtr, &Info, /* Allocator */ nullptr,
-                             /* Number of accessible agents (out) */ nullptr,
-                             /* Accessible agents */ nullptr);
+    hsa_status_t Status = hsa_amd_pointer_info(
+        HstPtr, &Info, /*Allocator=*/nullptr, /*num_agents_accessible=*/nullptr,
+        /*accessible=*/nullptr);
     if (auto Err = Plugin::check(Status, "Error in hsa_amd_pointer_info: %s"))
       return std::move(Err);
 
@@ -2789,7 +2788,7 @@ struct AMDGPUDeviceTy : public GenericDeviceTy, AMDGenericDeviceTy {
   AMDHostDeviceTy &HostDevice;
 
   /// The current size of the global device memory pool (managed by us).
-  uint64_t DeviceMemoryPoolSize = 1L << 29L /* 512MB */;
+  uint64_t DeviceMemoryPoolSize = 1L << 29L /*512MB=*/;
 
   /// The current size of the stack that will be used in cases where it could
   /// not be statically determined.
@@ -3031,9 +3030,8 @@ struct AMDGPUPluginTy final : public GenericPluginTy {
   /// Check whether the image is compatible with an AMDGPU device.
   Expected<bool> isELFCompatible(StringRef Image) const override {
     // Get the associated architecture and flags from the ELF.
-    auto ElfOrErr =
-        ELF64LEObjectFile::create(MemoryBufferRef(Image, /*Identifier=*/""),
-                                  /*InitContent=*/false);
+    auto ElfOrErr = ELF64LEObjectFile::create(
+        MemoryBufferRef(Image, /*Identifier=*/""), /*InitContent=*/false);
     if (!ElfOrErr)
       return ElfOrErr.takeError();
     std::optional<StringRef> Processor = ElfOrErr->tryGetCPUName();
diff --git a/openmp/libomptarget/plugins-nextgen/common/include/GlobalHandler.h b/openmp/libomptarget/plugins-nextgen/common/include/GlobalHandler.h
index d9fe938790ca76..8707e7b4c504e3 100644
--- a/openmp/libomptarget/plugins-nextgen/common/include/GlobalHandler.h
+++ b/openmp/libomptarget/plugins-nextgen/common/include/GlobalHandler.h
@@ -138,7 +138,7 @@ class GenericGlobalHandlerTy {
                              const GlobalTy &HostGlobal,
                              const GlobalTy &DeviceGlobal) {
     return moveGlobalBetweenDeviceAndHost(Device, HostGlobal, DeviceGlobal,
-                                          /* D2H */ true);
+                                          /*D2H=*/true);
   }
 
   /// Copy the memory associated with a global from the device to its
@@ -147,7 +147,7 @@ class GenericGlobalHandlerTy {
   Error readGlobalFromDevice(GenericDeviceTy &Device, DeviceImageTy &Image,
                              const GlobalTy &HostGlobal) {
     return moveGlobalBetweenDeviceAndHost(Device, Image, HostGlobal,
-                                          /* D2H */ true);
+                                          /*D2H=*/true);
   }
 
   /// Copy the memory associated with a global from the host to its counterpart
@@ -156,7 +156,7 @@ class GenericGlobalHandlerTy {
   Error writeGlobalToDevice(GenericDeviceTy &Device, const GlobalTy &HostGlobal,
                             const GlobalTy &DeviceGlobal) {
     return moveGlobalBetweenDeviceAndHost(Device, HostGlobal, DeviceGlobal,
-                                          /* D2H */ false);
+                                          /*D2H=*/false);
   }
 
   /// Copy the memory associated with a global from the host to its counterpart
@@ -165,7 +165,7 @@ class GenericGlobalHandlerTy {
   Error writeGlobalToDevice(GenericDeviceTy &Device, DeviceImageTy &Image,
                             const GlobalTy &HostGlobal) {
     return moveGlobalBetweenDeviceAndHost(Device, Image, HostGlobal,
-                                          /* D2H */ false);
+                                          /*D2H=*/false);
   }
 };
 
diff --git a/openmp/libomptarget/plugins-nextgen/common/src/JIT.cpp b/openmp/libomptarget/plugins-nextgen/common/src/JIT.cpp
index 7275be4edfca5b..9eb610cab4de66 100644
--- a/openmp/libomptarget/plugins-nextgen/common/src/JIT.cpp
+++ b/openmp/libomptarget/plugins-nextgen/common/src/JIT.cpp
@@ -93,7 +93,7 @@ createModuleFromImage(const __tgt_device_image &Image, LLVMContext &Context) {
   StringRef Data((const char *)Image.ImageStart,
                  target::getPtrDiff(Image.ImageEnd, Image.ImageStart));
   std::unique_ptr<MemoryBuffer> MB = MemoryBuffer::getMemBuffer(
-      Data, /* BufferName */ "", /* RequiresNullTerminator */ false);
+      Data, /*BufferName=*/"", /*RequiresNullTerminator=*/false);
   return createModuleFromMemoryBuffer(MB, Context);
 }
 
@@ -186,7 +186,7 @@ void JITEngine::codegen(TargetMachine *TM, TargetLibraryInfoImpl *TLII,
   TM->addPassesToEmitFile(PM, OS, nullptr,
                           TT.isNVPTX() ? CodeGenFileType::AssemblyFile
                                        : CodeGenFileType::ObjectFile,
-                          /* DisableVerify */ false, MMIWP);
+                          /*DisableVerify=*/false, MMIWP);
 
   PM.run(M);
 }
@@ -196,8 +196,8 @@ JITEngine::backend(Module &M, const std::string &ComputeUnitKind,
                    unsigned OptLevel) {
 
   auto RemarksFileOrErr = setupLLVMOptimizationRemarks(
-      M.getContext(), /* RemarksFilename */ "", /* RemarksPasses */ "",
-      /* RemarksFormat */ "", /* RemarksWithHotness */ false);
+      M.getContext(), /*RemarksFilename=*/"", /*RemarksPasses=*/"",
+      /*RemarksFormat=*/"", /*RemarksWithHotness=*/false);
   if (Error E = RemarksFileOrErr.takeError())
     return std::move(E);
   if (*RemarksFileOrErr)
diff --git a/openmp/libomptarget/plugins-nextgen/common/src/PluginInterface.cpp b/openmp/libomptarget/plugins-nextgen/common/src/PluginInterface.cpp
index 9490e58fc669cd..0db7910ec10567 100644
--- a/openmp/libomptarget/plugins-nextgen/common/src/PluginInterface.cpp
+++ b/openmp/libomptarget/plugins-nextgen/common/src/PluginInterface.cpp
@@ -64,7 +64,7 @@ struct RecordReplayTy {
   void *suggestAddress(uint64_t MaxMemoryAllocation) {
     // Get a valid pointer address for this system
     void *Addr =
-        Device->allocate(1024, /* HstPtr */ nullptr, TARGET_ALLOC_DEFAULT);
+        Device->allocate(1024, /*HstPtr=*/nullptr, TARGET_ALLOC_DEFAULT);
     Device->free(Addr);
     // Align Address to MaxMemoryAllocation
     Addr = (void *)alignPtr((Addr), MaxMemoryAllocation);
@@ -104,8 +104,8 @@ struct RecordReplayTy {
     constexpr size_t STEP = 1024 * 1024 * 1024ULL;
     MemoryStart = nullptr;
     for (TotalSize = MAX_MEMORY_ALLOCATION; TotalSize > 0; TotalSize -= STEP) {
-      MemoryStart = Device->allocate(TotalSize, /* HstPtr */ nullptr,
-                                     TARGET_ALLOC_DEFAULT);
+      MemoryStart =
+          Device->allocate(TotalSize, /*HstPtr=*/nullptr, TARGET_ALLOC_DEFAULT);
       if (MemoryStart)
         break;
     }
@@ -214,8 +214,8 @@ struct RecordReplayTy {
     for (auto &OffloadEntry : Image.getOffloadEntryTable()) {
       if (!OffloadEntry.size)
         continue;
-      Size += std::strlen(OffloadEntry.name) + /* '\0' */ 1 +
-              /* OffloadEntry.size value */ sizeof(uint32_t) +
+      // Get the total size of the string and entry including the null byte.
+      Size += std::strlen(OffloadEntry.name) + 1 + sizeof(uint32_t) +
               OffloadEntry.size;
     }
 
@@ -735,13 +735,12 @@ Error GenericDeviceTy::init(GenericPluginTy &Plugin) {
   if (ompt::Initialized) {
     bool ExpectedStatus = false;
     if (OmptInitialized.compare_exchange_strong(ExpectedStatus, true))
-      performOmptCallback(device_initialize,
-                          /* device_num */ DeviceId +
-                              Plugin.getDeviceIdStartIndex(),
-                          /* type */ getComputeUnitKind().c_str(),
-                          /* device */ reinterpret_cast<ompt_device_t *>(this),
-                          /* lookup */ ompt::lookupCallbackByName,
-                          /* documentation */ nullptr);
+      performOmptCallback(device_initialize, /*device_num=*/DeviceId +
+                                                 Plugin.getDeviceIdStartIndex(),
+                          /*type=*/getComputeUnitKind().c_str(),
+                          /*device=*/reinterpret_cast<ompt_device_t *>(this),
+                          /*lookup=*/ompt::lookupCallbackByName,
+                          /*documentation=*/nullptr);
   }
 #endif
 
@@ -835,7 +834,7 @@ Error GenericDeviceTy::deinit(GenericPluginTy &Plugin) {
     bool ExpectedStatus = true;
     if (OmptInitialized.compare_exchange_strong(ExpectedStatus, false))
       performOmptCallback(device_finalize,
-                          /* device_num */ DeviceId +
+                          /*device_num=*/DeviceId +
                               Plugin.getDeviceIdStartIndex());
   }
 #endif
@@ -897,16 +896,11 @@ GenericDeviceTy::loadBinary(GenericPluginTy &Plugin,
   if (ompt::Initialized) {
     size_t Bytes =
         getPtrDiff(InputTgtImage->ImageEnd, InputTgtImage->ImageStart);
-    performOmptCallback(device_load,
-                        /* device_num */ DeviceId +
-                            Plugin.getDeviceIdStartIndex(),
-                        /* FileName */ nullptr,
-                        /* File Offset */ 0,
-                        /* VmaInFile */ nullptr,
-                        /* ImgSize */ Bytes,
-                        /* HostAddr */ InputTgtImage->ImageStart,
-                        /* DeviceAddr */ nullptr,
-                        /* FIXME: ModuleId */ 0);
+    performOmptCallback(
+        device_load, /*device_num=*/DeviceId + Plugin.getDeviceIdStartIndex(),
+        /*FileName=*/nullptr, /*FileOffset=*/0, /*VmaInFile=*/nullptr,
+        /*ImgSize=*/Bytes, /*HostAddr=*/InputTgtImage->ImageStart,
+        /*DeviceAddr=*/nullptr, /* FIXME: ModuleId */ 0);
   }
 #endif
 
@@ -1293,7 +1287,7 @@ Error PinnedAllocationMapTy::lockMappedHostBuffer(void *HstPtr, size_t Size) {
   // If pinned, just insert the entry representing the whole pinned buffer.
   if (*IsPinnedOrErr)
     return insertEntry(BaseHstPtr, BaseDevAccessiblePtr, BaseSize,
-                       /* Externally locked */ true);
+                       /*Externallylocked=*/true);
 
   // Not externally pinned. Do nothing if locking of mapped buffers is disabled.
   if (!LockMappedBuffers)
@@ -1863,7 +1857,7 @@ int32_t __tgt_rtl_data_notify_unmapped(int32_t DeviceId, void *HstPtr) {
 int32_t __tgt_rtl_data_submit(int32_t DeviceId, void *TgtPtr, void *HstPtr,
                               int64_t Size) {
   return __tgt_rtl_data_submit_async(DeviceId, TgtPtr, HstPtr, Size,
-                                     /* AsyncInfoPtr */ nullptr);
+                                     /*AsyncInfoPtr=*/nullptr);
 }
 
 int32_t __tgt_rtl_data_submit_async(int32_t DeviceId, void *TgtPtr,
@@ -1885,7 +1879,7 @@ int32_t __tgt_rtl_data_submit_async(int32_t DeviceId, void *TgtPtr,
 int32_t __tgt_rtl_data_retrieve(int32_t DeviceId, void *HstPtr, void *TgtPtr,
                                 int64_t Size) {
   return __tgt_rtl_data_retrieve_async(DeviceId, HstPtr, TgtPtr, Size,
-                                       /* AsyncInfoPtr */ nullptr);
+                                       /*AsyncInfoPtr=*/nullptr);
 }
 
 int32_t __tgt_rtl_data_retrieve_async(int32_t DeviceId, void *HstPtr,
@@ -1909,7 +1903,7 @@ int32_t __tgt_rtl_data_exchange(int32_t SrcDeviceId, void *SrcPtr,
                                 int64_t Size) {
   return __tgt_rtl_data_exchange_async(SrcDeviceId, SrcPtr, DstDeviceId, DstPtr,
                                        Size,
-                                       /* AsyncInfoPtr */ nullptr);
+                                       /*AsyncInfoPtr=*/nullptr);
 }
 
 int32_t __tgt_rtl_data_exchange_async(int32_t SrcDeviceId, void *SrcPtr,
diff --git a/openmp/libomptarget/plugins-nextgen/cuda/src/rtl.cpp b/openmp/libomptarget/plugins-nextgen/cuda/src/rtl.cpp
index 0005bff7a8035d..fb51f96c07b74d 100644
--- a/openmp/libomptarget/plugins-nextgen/cuda/src/rtl.cpp
+++ b/openmp/libomptarget/plugins-nextgen/cuda/src/rtl.cpp
@@ -1216,10 +1216,10 @@ Error CUDAKernelTy::launchImpl(GenericDeviceTy &GenericDevice,
       std::max(KernelArgs.DynCGroupMem, GenericDevice.getDynamicMemorySize());
 
   CUresult Res =
-      cuLaunchKernel(Func, NumBlocks, /* gridDimY */ 1,
-                     /* gridDimZ */ 1, NumThreads,
-                     /* blockDimY */ 1, /* blockDimZ */ 1, MaxDynCGroupMem,
-                     Stream, (void **)Args, nullptr);
+      cuLaunchKernel(Func, NumBlocks, /*gridDimY=*/1,
+                     /*gridDimZ=*/1, NumThreads,
+                     /*blockDimY=*/1, /*blockDimZ=*/1, MaxDynCGroupMem, Stream,
+                     (void **)Args, nullptr);
   return Plugin::check(Res, "Error in cuLaunchKernel for '%s': %s", getName());
 }
 
diff --git a/openmp/libomptarget/plugins-nextgen/generic-elf-64bit/src/rtl.cpp b/openmp/libomptarget/plugins-nextgen/generic-elf-64bit/src/rtl.cpp
index 7f66b6827bce00..6466afc543b568 100644
--- a/openmp/libomptarget/plugins-nextgen/generic-elf-64bit/src/rtl.cpp
+++ b/openmp/libomptarget/plugins-nextgen/generic-elf-64bit/src/rtl.cpp
@@ -73,8 +73,8 @@ struct GenELF64KernelTy : public GenericKernelTy {
     Func = (void (*)())Global.getPtr();
 
     KernelEnvironment.Configuration.ExecMode = OMP_TGT_EXEC_MODE_GENERIC;
-    KernelEnvironment.Configuration.MayUseNestedParallelism = /* Unknown */ 2;
-    KernelEnvironment.Configuration.UseGenericStateMachine = /* Unknown */ 2;
+    KernelEnvironment.Configuration.MayUseNestedParallelism = /*Unknown=*/2;
+    KernelEnvironment.Configuration.UseGenericStateMachine = /*Unknown=*/2;
 
     // Set the maximum number of threads to a single.
     MaxNumThreads = 1;
diff --git a/openmp/libomptarget/src/OpenMP/Mapping.cpp b/openmp/libomptarget/src/OpenMP/Mapping.cpp
index a5c24810e0af95..833856f2abf290 100644
--- a/openmp/libomptarget/src/OpenMP/Mapping.cpp
+++ b/openmp/libomptarget/src/OpenMP/Mapping.cpp
@@ -303,9 +303,9 @@ TargetPointerResultTy MappingInfoTy::getTargetPointer(
 
     // Notify the plugin about the new mapping.
     if (Device.notifyDataMapped(HstPtrBegin, Size))
-      return {{false /* IsNewEntry */, false /* IsHostPointer */},
-              nullptr /* Entry */,
-              nullptr /* TargetPointer */};
+      return {{false /*IsNewEntry=*/, false /*IsHostPointer=*/},
+              nullptr /*Entry=*/,
+              nullptr /*TargetPointer=*/};
   } else {
     // This entry is not present and we did not create a new entry for it.
     LR.TPR.Flags.IsPresent = false;
@@ -333,9 +333,9 @@ TargetPointerResultTy MappingInfoTy::getTargetPointer(
       LR.TPR.TargetPointer = nullptr;
     } else if (LR.TPR.getEntry()->addEventIfNecessary(Device, AsyncInfo) !=
                OFFLOAD_SUCCESS)
-      return {{false /* IsNewEntry */, false /* IsHostPointer */},
-              nullptr /* Entry */,
-              nullptr /* TargetPointer */};
+      return {{false /*IsNewEntry=*/, false /*IsHostPointer=*/},
+              nullptr /*Entry=*/,
+              nullptr /*TargetPointer=*/};
   } else {
     // If not a host pointer and no present modifier, we need to wait for the
     // event if it exists.
@@ -349,9 +349,9 @@ TargetPointerResultTy MappingInfoTy::getTargetPointer(
           // If it fails to wait for the event, we need to return nullptr in
           // case of any data race.
           REPORT("Failed to wait for event " DPxMOD ".\n", DPxPTR(Event));
-          return {{false /* IsNewEntry */, false /* IsHostPointer */},
-                  nullptr /* Entry */,
-                  nullptr /* TargetPointer */};
+          return {{false /*IsNewEntry=*/, false /*IsHostPointer=*/},
+                  nullptr /*Entry=*/,
+                  nullptr /*TargetPointer=*/};
         }
       }
     }
diff --git a/openmp/libomptarget/src/OpenMP/OMPT/Callback.cpp b/openmp/libomptarget/src/OpenMP/OMPT/Callback.cpp
index da955e101956f4..82934f10486c5a 100644
--- a/openmp/libomptarget/src/OpenMP/OMPT/Callback.cpp
+++ b/openmp/libomptarget/src/OpenMP/OMPT/Callback.cpp
@@ -88,16 +88,16 @@ void Interface::beginTargetDataAlloc(int64_t DeviceId, void *HstPtrBegin,
     ompt_callback_target_data_op_emi_fn(
         ompt_scope_begin, TargetTaskData, &TargetData, &HostOpId,
         ompt_target_data_alloc, HstPtrBegin,
-        /* SrcDeviceNum */ omp_get_initial_device(), *TgtPtrBegin,
-        /* TgtDeviceNum */ DeviceId, Size, Code);
+        /*SrcDeviceNum=*/omp_get_initial_device(), *TgtPtrBegin,
+        /*TgtDeviceNum=*/DeviceId, Size, Code);
   } else if (ompt_callback_target_data_op_fn) {
     // HostOpId is set by the runtime
     HostOpId = createOpId();
     // Invoke the tool supplied data op callback
     ompt_callback_target_data_op_fn(
         TargetData.value, HostOpId, ompt_target_data_alloc, HstPtrBegin,
-        /* SrcDeviceNum */ omp_get_initial_device(), *TgtPtrBegin,
-        /* TgtDeviceNum */ DeviceId, Size, Code);
+        /*SrcDeviceNum=*/omp_get_initial_device(), *TgtPtrBegin,
+        /*TgtDeviceNum=*/DeviceId, Size, Code);
   }
 }
 
@@ -111,8 +111,8 @@ void Interface::endTargetDataAlloc(int64_t DeviceId, void *HstPtrBegin,
     ompt_callback_target_data_op_emi_fn(
         ompt_scope_end, TargetTaskData, &TargetData, &HostOpId,
         ompt_target_data_alloc, HstPtrBegin,
-        /* SrcDeviceNum */ omp_get_initial_device(), *TgtPtrBegin,
-        /* TgtDeviceNum */ DeviceId, Size, Code);
+        /*SrcDeviceNum=*/omp_get_initial_device(), *TgtPtrBegin,
+        /*TgtDeviceNum=*/DeviceId, Size, Code);
   }
   endTargetDataOperation();
 }
@@ -127,15 +127,15 @@ void Interface::beginTargetDataSubmit(int64_t DeviceId, void *TgtPtrBegin,
     ompt_callback_target_data_op_emi_fn(
         ompt_scope_begin, TargetTaskData, &TargetData, &HostOpId,
         ompt_target_data_transfer_to_device, HstPtrBegin,
-        /* SrcDeviceNum */ omp_get_initial_device(), TgtPtrBegin, DeviceId,
-        Size, Code);
+        /*SrcDeviceNum=*/omp_get_initial_device(), TgtPtrBegin, DeviceId, Size,
+        Code);
   } else if (ompt_callback_target_data_op_fn) {
     // HostOpId is set by the runtime
     HostOpId = createOpId();
     // Invoke the tool supplied data op callback
     ompt_callback_target_data_op_fn(
         TargetData.value, HostOpId, ompt_target_data_transfer_to_device,
-        HstPtrBegin, /* SrcDeviceNum */ omp_get_initial_device(), TgtPtrBegin,
+        HstPtrBegin, /*SrcDeviceNum=*/omp_get_initial_device(), TgtPtrBegin,
         DeviceId, Size, Code);
   }
 }
@@ -150,8 +150,8 @@ void Interface::endTargetDataSubmit(int64_t DeviceId, void *TgtPtrBegin,
     ompt_callback_target_data_op_emi_fn(
         ompt_scope_end, TargetTaskData, &TargetData, &HostOpId,
         ompt_target_data_transfer_to_device, HstPtrBegin,
-        /* SrcDeviceNum */ omp_get_initial_device(), TgtPtrBegin, DeviceId,
-        Size, Code);
+        /*SrcDeviceNum=*/omp_get_initial_device(), TgtPtrBegin, DeviceId, Size,
+        Code);
   }
   endTargetDataOperation();
 }
@@ -165,15 +165,15 @@ void Interface::beginTargetDataDelete(int64_t DeviceId, void *TgtPtrBegin,
     ompt_callback_target_data_op_emi_fn(
         ompt_scope_begin, TargetTaskData, &TargetData, &HostOpId,
         ompt_target_data_delete, TgtPtrBegin, DeviceId,
-        /* TgtPtrBegin */ nullptr, /* TgtDeviceNum */ -1, /* Bytes */ 0, Code);
+        /*TgtPtrBegin=*/nullptr, /*TgtDeviceNum=*/-1, /*Bytes=*/0, Code);
   } else if (ompt_callback_target_data_op_fn) {
     // HostOpId is set by the runtime
     HostOpId = createOpId();
     // Invoke the tool supplied data op callback
     ompt_callback_target_data_op_fn(TargetData.value, HostOpId,
                                     ompt_target_data_delete, TgtPtrBegin,
-                                    DeviceId, /* TgtPtrBegin */ nullptr,
-                                    /* TgtDeviceNum */ -1, /* Bytes */ 0, Code);
+                                    DeviceId, /*TgtPtrBegin=*/nullptr,
+                                    /*TgtDeviceNum=*/-1, /*Bytes=*/0, Code);
   }
 }
 
@@ -186,7 +186,7 @@ void Interface::endTargetDataDelete(int64_t DeviceId, void *TgtPtrBegin,
     ompt_callback_target_data_op_emi_fn(
         ompt_scope_end, TargetTaskData, &TargetData, &HostOpId,
         ompt_target_data_delete, TgtPtrBegin, DeviceId,
-        /* TgtPtrBegin */ nullptr, /* TgtDeviceNum */ -1, /* Bytes */ 0, Code);
+        /*TgtPtrBegin=*/nullptr, /*TgtDeviceNum=*/-1, /*Bytes=*/0, Code);
   }
   endTargetDataOperation();
 }
@@ -202,7 +202,7 @@ void Interface::beginTargetDataRetrieve(int64_t DeviceId, void *HstPtrBegin,
         ompt_scope_begin, TargetTaskData, &TargetData, &HostOpId,
         ompt_target_data_transfer_from_device, TgtPtrBegin, DeviceId,
         HstPtrBegin,
-        /* TgtDeviceNum */ omp_get_initial_device(), Size, Code);
+        /*TgtDeviceNum=*/omp_get_initial_device(), Size, Code);
   } else if (ompt_callback_target_data_op_fn) {
     // HostOpId is set by the runtime
     HostOpId = createOpId();
@@ -210,7 +210,7 @@ void Interface::beginTargetDataRetrieve(int64_t DeviceId, void *HstPtrBegin,
     ompt_callback_target_data_op_fn(
         TargetData.value, HostOpId, ompt_target_data_transfer_from_device,
         TgtPtrBegin, DeviceId, HstPtrBegin,
-        /* TgtDeviceNum */ omp_get_initial_device(), Size, Code);
+        /*TgtDeviceNum=*/omp_get_initial_device(), Size, Code);
   }
 }
 
@@ -225,7 +225,7 @@ void Interface::endTargetDataRetrieve(int64_t DeviceId, void *HstPtrBegin,
         ompt_scope_end, TargetTaskData, &TargetData, &HostOpId,
         ompt_target_data_transfer_from_device, TgtPtrBegin, DeviceId,
         HstPtrBegin,
-        /* TgtDeviceNum */ omp_get_initial_device(), Size, Code);
+        /*TgtDeviceNum=*/omp_get_initial_device(), Size, Code);
   }
   endTargetDataOperation();
 }
diff --git a/openmp/libomptarget/src/device.cpp b/openmp/libomptarget/src/device.cpp
index fa8932361a51c9..654efd524024a4 100644
--- a/openmp/libomptarget/src/device.cpp
+++ b/openmp/libomptarget/src/device.cpp
@@ -117,7 +117,7 @@ void *DeviceTy::allocData(int64_t Size, void *HstPtr, int32_t Kind) {
   OMPT_IF_BUILT(InterfaceRAII TargetDataAllocRAII(
                     RegionInterface.getCallbacks<ompt_target_data_alloc>(),
                     DeviceID, HstPtr, &TargetPtr, Size,
-                    /* CodePtr */ OMPT_GET_RETURN_ADDRESS(0));)
+                    /*CodePtr=*/OMPT_GET_RETURN_ADDRESS(0));)
 
   TargetPtr = RTL->data_alloc(RTLDeviceID, Size, HstPtr, Kind);
   return TargetPtr;
@@ -128,7 +128,7 @@ int32_t DeviceTy::deleteData(void *TgtAllocBegin, int32_t Kind) {
   OMPT_IF_BUILT(InterfaceRAII TargetDataDeleteRAII(
                     RegionInterface.getCallbacks<ompt_target_data_delete>(),
                     DeviceID, TgtAllocBegin,
-                    /* CodePtr */ OMPT_GET_RETURN_ADDRESS(0));)
+                    /*CodePtr=*/OMPT_GET_RETURN_ADDRESS(0));)
 
   return RTL->data_delete(RTLDeviceID, TgtAllocBegin, Kind);
 }
@@ -146,7 +146,7 @@ int32_t DeviceTy::submitData(void *TgtPtrBegin, void *HstPtrBegin, int64_t Size,
       InterfaceRAII TargetDataSubmitRAII(
           RegionInterface.getCallbacks<ompt_target_data_transfer_to_device>(),
           DeviceID, TgtPtrBegin, HstPtrBegin, Size,
-          /* CodePtr */ OMPT_GET_RETURN_ADDRESS(0));)
+          /*CodePtr=*/OMPT_GET_RETURN_ADDRESS(0));)
 
   if (!AsyncInfo || !RTL->data_submit_async || !RTL->synchronize)
     return RTL->data_submit(RTLDeviceID, TgtPtrBegin, HstPtrBegin, Size);
@@ -168,7 +168,7 @@ int32_t DeviceTy::retrieveData(void *HstPtrBegin, void *TgtPtrBegin,
       InterfaceRAII TargetDataRetrieveRAII(
           RegionInterface.getCallbacks<ompt_target_data_transfer_from_device>(),
           DeviceID, HstPtrBegin, TgtPtrBegin, Size,
-          /* CodePtr */ OMPT_GET_RETURN_ADDRESS(0));)
+          /*CodePtr=*/OMPT_GET_RETURN_ADDRESS(0));)
 
   if (!RTL->data_retrieve_async || !RTL->synchronize)
     return RTL->data_retrieve(RTLDeviceID, HstPtrBegin, TgtPtrBegin, Size);
diff --git a/openmp/libomptarget/src/interface.cpp b/openmp/libomptarget/src/interface.cpp
index 61d9db17f51006..49495ac266f1b3 100644
--- a/openmp/libomptarget/src/interface.cpp
+++ b/openmp/libomptarget/src/interface.cpp
@@ -113,7 +113,7 @@ targetData(ident_t *Loc, int64_t DeviceId, int32_t ArgNum, void **ArgsBase,
   int Rc = OFFLOAD_SUCCESS;
   Rc = TargetDataFunction(Loc, *DeviceOrErr, ArgNum, ArgsBase, Args, ArgSizes,
                           ArgTypes, ArgNames, ArgMappers, AsyncInfo,
-                          false /* FromMapper */);
+                          false /*FromMapper=*/);
 
   if (Rc == OFFLOAD_SUCCESS)
     Rc = AsyncInfo.synchronize();
@@ -293,7 +293,7 @@ static inline int targetKernel(ident_t *Loc, int64_t DeviceId, int32_t NumTeams,
   /// RAII to establish tool anchors before and after target region
   OMPT_IF_BUILT(InterfaceRAII TargetRAII(
                     RegionInterface.getCallbacks<ompt_target>(), DeviceId,
-                    /* CodePtr */ OMPT_GET_RETURN_ADDRESS(0));)
+                    /*CodePtr=*/OMPT_GET_RETURN_ADDRESS(0));)
 
   int Rc = OFFLOAD_SUCCESS;
   Rc = target(Loc, *DeviceOrErr, HostPtr, *KernelArgs, AsyncInfo);
@@ -387,7 +387,7 @@ EXTERN int __tgt_target_kernel_replay(ident_t *Loc, int64_t DeviceId,
   /// RAII to establish tool anchors before and after target region
   OMPT_IF_BUILT(InterfaceRAII TargetRAII(
                     RegionInterface.getCallbacks<ompt_target>(), DeviceId,
-                    /* CodePtr */ OMPT_GET_RETURN_ADDRESS(0));)
+                    /*CodePtr=*/OMPT_GET_RETURN_ADDRESS(0));)
 
   AsyncInfoTy AsyncInfo(*DeviceOrErr);
   int Rc = target_replay(Loc, *DeviceOrErr, HostPtr, DeviceMemory,
diff --git a/openmp/libomptarget/src/omptarget.cpp b/openmp/libomptarget/src/omptarget.cpp
index a7d55d7ebd5391..ed34870e12d011 100644
--- a/openmp/libomptarget/src/omptarget.cpp
+++ b/openmp/libomptarget/src/omptarget.cpp
@@ -225,7 +225,7 @@ static int initLibrary(DeviceTy &Device) {
           AsyncInfoTy AsyncInfo(Device);
           void *DevPtr;
           Device.retrieveData(&DevPtr, CurrDeviceEntryAddr, sizeof(void *),
-                              AsyncInfo, /* Entry */ nullptr, &HDTTMap);
+                              AsyncInfo, /*Entry=*/nullptr, &HDTTMap);
           if (AsyncInfo.synchronize() != OFFLOAD_SUCCESS)
             return OFFLOAD_FAIL;
           CurrDeviceEntryAddr = DevPtr;
@@ -617,7 +617,7 @@ int targetDataBegin(ident_t *Loc, DeviceTy &Device, int32_t ArgNum,
           /*HstPtrName=*/nullptr,
           /*HasFlagTo=*/false, /*HasFlagAlways=*/false, IsImplicit, UpdateRef,
           HasCloseModifier, HasPresentModifier, HasHoldModifier, AsyncInfo,
-          /* OwnedTPR */ nullptr, /* ReleaseHDTTMap */ false);
+          /*OwnedTPR=*/nullptr, /*ReleaseHDTTMap=*/false);
       PointerTgtPtrBegin = PointerTpr.TargetPointer;
       IsHostPtr = PointerTpr.Flags.IsHostPointer;
       if (!PointerTgtPtrBegin) {
@@ -1723,7 +1723,7 @@ int target_replay(ident_t *Loc, DeviceTy &Device, void *HostPtr,
   DP("Launching target execution %s with pointer " DPxMOD " (index=%d).\n",
      TargetTable->EntriesBegin[TM->Index].name, DPxPTR(TgtEntryPtr), TM->Index);
 
-  void *TgtPtr = Device.allocData(DeviceMemorySize, /* HstPtr */ nullptr,
+  void *TgtPtr = Device.allocData(DeviceMemorySize, /*HstPtr=*/nullptr,
                                   TARGET_ALLOC_DEFAULT);
   Device.submitData(TgtPtr, DeviceMemory, DeviceMemorySize, AsyncInfo);
 
diff --git a/openmp/libomptarget/tools/kernelreplay/llvm-omp-kernel-replay.cpp b/openmp/libomptarget/tools/kernelreplay/llvm-omp-kernel-replay.cpp
index fc3ff078a81fb9..761e04e4c7bbdb 100644
--- a/openmp/libomptarget/tools/kernelreplay/llvm-omp-kernel-replay.cpp
+++ b/openmp/libomptarget/tools/kernelreplay/llvm-omp-kernel-replay.cpp
@@ -57,8 +57,8 @@ int main(int argc, char **argv) {
   cl::ParseCommandLineOptions(argc, argv, "llvm-omp-kernel-replay\n");
 
   ErrorOr<std::unique_ptr<MemoryBuffer>> KernelInfoMB =
-      MemoryBuffer::getFile(InputFilename, /* isText */ true,
-                            /* RequiresNullTerminator */ true);
+      MemoryBuffer::getFile(InputFilename, /*isText=*/true,
+                            /*RequiresNullTerminator=*/true);
   if (!KernelInfoMB)
     report_fatal_error("Error reading the kernel info json file");
   Expected<json::Value> JsonKernelInfo =
@@ -100,8 +100,8 @@ int main(int argc, char **argv) {
   KernelEntry.addr = (void *)0x1;
 
   ErrorOr<std::unique_ptr<MemoryBuffer>> ImageMB =
-      MemoryBuffer::getFile(KernelEntryName + ".image", /* isText */ false,
-                            /* RequiresNullTerminator */ false);
+      MemoryBuffer::getFile(KernelEntryName + ".image", /*isText=*/false,
+                            /*RequiresNullTerminator=*/false);
   if (!ImageMB)
     report_fatal_error("Error reading the kernel image.");
 
@@ -127,7 +127,7 @@ int main(int argc, char **argv) {
   int32_t DeviceId = (DeviceIdOpt > -1 ? DeviceIdOpt : DeviceIdJson.value());
 
   // TODO: do we need requires?
-  //__tgt_register_requires(/* Flags */1);
+  //__tgt_register_requires(/*Flags=*/1);
 
   __tgt_register_lib(&Desc);
 
@@ -140,8 +140,8 @@ int main(int argc, char **argv) {
   }
 
   ErrorOr<std::unique_ptr<MemoryBuffer>> DeviceMemoryMB =
-      MemoryBuffer::getFile(KernelEntryName + ".memory", /* isText */ false,
-                            /* RequiresNullTerminator */ false);
+      MemoryBuffer::getFile(KernelEntryName + ".memory", /*isText=*/false,
+                            /*RequiresNullTerminator=*/false);
 
   if (!DeviceMemoryMB)
     report_fatal_error("Error reading the kernel input device memory.");
@@ -166,7 +166,7 @@ int main(int argc, char **argv) {
   }
 
   __tgt_target_kernel_replay(
-      /* Loc */ nullptr, DeviceId, KernelEntry.addr, (char *)recored_data,
+      /*Loc=*/nullptr, DeviceId, KernelEntry.addr, (char *)recored_data,
       DeviceMemoryMB.get()->getBufferSize(), TgtArgs.data(),
       TgtArgOffsets.data(), NumArgs.value(), NumTeams, NumThreads,
       LoopTripCount.value());
@@ -174,15 +174,15 @@ int main(int argc, char **argv) {
   if (VerifyOpt) {
     ErrorOr<std::unique_ptr<MemoryBuffer>> OriginalOutputMB =
         MemoryBuffer::getFile(KernelEntryName + ".original.output",
-                              /* isText */ false,
-                              /* RequiresNullTerminator */ false);
+                              /*isText=*/false,
+                              /*RequiresNullTerminator=*/false);
     if (!OriginalOutputMB)
       report_fatal_error("Error reading the kernel original output file, make "
                          "sure LIBOMPTARGET_SAVE_OUTPUT is set when recording");
     ErrorOr<std::unique_ptr<MemoryBuffer>> ReplayOutputMB =
         MemoryBuffer::getFile(KernelEntryName + ".replay.output",
-                              /* isText */ false,
-                              /* RequiresNullTerminator */ false);
+                              /*isText=*/false,
+                              /*RequiresNullTerminator=*/false);
     if (!ReplayOutputMB)
       report_fatal_error("Error reading the kernel replay output file");
 



More information about the Openmp-commits mailing list