Address review comments

Author: 0x12CC

clang/lib/Driver/ToolChains/SYCL.cpp

@@ -1574,6 +1574,7 @@ static ArrayRef<options::ID> getUnsupportedOpts() {
       options::OPT_fno_profile_generate, // -f[no-]profile-generate
       options::OPT_ftest_coverage,
       options::OPT_fno_test_coverage, // -f[no-]test-coverage
+      options::OPT_coverage,          // --coverage
       options::OPT_fprofile_arcs,
       options::OPT_fno_profile_arcs,           // -f[no-]profile-arcs
       options::OPT_fno_profile_instr_generate, // -f[no-]profile-instr-generate

clang/test/Driver/sycl-unsupported.cpp

@@ -19,13 +19,6 @@
 // RUN:    -DOPT_CC1=-debug-info-kind=line-tables-only \
 // RUN:    -check-prefixes=UNSUPPORTED_OPT_DIAG,UNSUPPORTED_OPT
 
-// RUN: %clangxx -fsycl -fprofile-instr-generate -### %s 2>&1 \
-// RUN:  | FileCheck %s -DARCH=spir64 -DOPT=-fprofile-instr-generate \
-// RUN:    -DOPT_CC1=-fprofile-instrument=clang \
-// RUN:    -check-prefixes=UNSUPPORTED_OPT_DIAG,UNSUPPORTED_OPT
-// RUN: %clangxx -fsycl -fcoverage-mapping \
-// RUN:          -fprofile-instr-generate -### %s 2>&1 \
-// RUN:  | FileCheck %s -DARCH=spir64 -DOPT=-fcoverage-mapping
 // RUN: %clangxx -fsycl -ftest-coverage -### %s 2>&1 \
 // RUN:  | FileCheck %s -DARCH=spir64 -DOPT=-ftest-coverage \
 // RUN:    -DOPT_CC1=-coverage-notes-file \
@@ -49,12 +42,6 @@
 // RUN:  | FileCheck %s -DARCH=spir64 -DOPT=--coverage \
 // RUN:    -DOPT_CC1=-coverage-notes-file \
 // RUN:    -check-prefixes=UNSUPPORTED_OPT_DIAG,UNSUPPORTED_OPT
-// Check to make sure our '-fsanitize=address' exception isn't triggered by a
-// different option
-// RUN: %clangxx -fsycl -fprofile-instr-generate=address -### %s 2>&1 \
-// RUN:  | FileCheck %s -DARCH=spir64 -DOPT=-fprofile-instr-generate=address \
-// RUN:    -DOPT_CC1=-fprofile-instrument=clang \
-// RUN:    -check-prefixes=UNSUPPORTED_OPT_DIAG,UNSUPPORTED_OPT
 
 // CHECK: ignoring '[[OPT]]' option as it is not currently supported for target '[[ARCH]]{{.*}}'; only supported for host compilation [-Woption-ignored]
 // CHECK-NOT: clang{{.*}} "-fsycl-is-device"{{.*}} "[[OPT]]{{.*}}"

llvm/lib/Transforms/Instrumentation/InstrProfiling.cpp

@@ -1126,9 +1126,10 @@ Value *InstrLowerer::getCounterAddress(InstrProfCntrInstBase *I) {
                                     Counters, "pgocount.addr");
     const std::uint64_t Index = I->getIndex()->getZExtValue();
     if (Index > 0) {
-      auto *Offset = Builder.getInt64(I->getIndex()->getZExtValue());
-      auto *AddrWithOffset = Builder.CreateGEP(Type::getInt64Ty(M.getContext()),
-                                               Addr, Offset, "pgocount.addr");
+      auto *Offset = Builder.getInt64(I->getIndex()->getZExtValue() *
+                                      sizeof(std::uint64_t));
+      auto *AddrWithOffset =
+          Builder.CreatePtrAdd(Addr, Offset, "pgocount.offset");
       return AddrWithOffset;
     }
     return Addr;
@@ -1686,12 +1687,7 @@ InstrLowerer::createRegionCounters(InstrProfCntrInstBase *Inc, StringRef Name,
                             Constant::getNullValue(StructTy), Name);
     const std::uint64_t FnHash = IndexedInstrProf::ComputeHash(
         getPGOFuncNameVarInitializer(Inc->getName()));
-    const std::string FnName = [&] {
-      auto *Arr = cast<ConstantDataArray>(Inc->getName()->getInitializer());
-      StringRef NameStr =
-          Arr->isCString() ? Arr->getAsCString() : Arr->getAsString();
-      return std::string{"__profc_"} + std::to_string(FnHash);
-    }();
+    const std::string FnName = std::string{"__profc_"} + std::to_string(FnHash);
     GV->addAttribute("sycl-unique-id", FnName);
     GV->addAttribute("sycl-device-global-size", Twine(NumCounters * 8).str());
     return GV;

sycl/source/detail/device_global_map_entry.cpp

@@ -79,22 +79,21 @@ void DeviceGlobalMapEntry::cleanupProfileCounter(context_impl *CtxImpl) {
     return std::strtoull(MUniqueId.substr(PrefixSize).c_str(), nullptr,
                          DecimalBase);
   }();
-  for (device_impl &Device : CtxImpl->getDevices()) {
+  for (const device_impl &Device : CtxImpl->getDevices()) {
     auto USMPtrIt = MDeviceToUSMPtrMap.find({&Device, CtxImpl});
     if (USMPtrIt != MDeviceToUSMPtrMap.end()) {
       DeviceGlobalUSMMem &USMMem = USMPtrIt->second;
 
-      // Get the increments from the USM pointer
+      // Get the increments from the USM pointer.
       std::vector<std::uint64_t> Increments(NumCounters);
       const std::uint64_t *Counters = static_cast<std::uint64_t *>(USMMem.MPtr);
       for (std::size_t I = 0; I < NumCounters; ++I)
-        Increments[I] += Counters[I];
+        Increments[I] = Counters[I];
 
-      // Call the weak symbol to update the profile counters
-      if (__sycl_increment_profile_counters) {
+      // Call the weak symbol to update the profile counters.
+      if (__sycl_increment_profile_counters)
         __sycl_increment_profile_counters(FnHash, Increments.size(),
                                           Increments.data());
-      }
 
       // Free the USM memory and release the event if it exists.
       detail::usm::freeInternal(USMMem.MPtr, CtxImpl);

sycl/source/detail/device_global_map_entry.hpp

@@ -116,7 +116,7 @@ struct DeviceGlobalMapEntry {
   // Returns true if the device_global is a profile counter.
   bool isProfileCounter();
 
-  // Cleans up a profile counter device global
+  // Cleans up a profile counter device global.
   void cleanupProfileCounter(context_impl *CtxImpl);
 
   // Gets or allocates USM memory for a device_global.

sycl/source/detail/program_manager/program_manager.cpp

@@ -2469,7 +2469,7 @@ std::vector<DeviceGlobalMapEntry *> ProgramManager::getDeviceGlobalEntries(
 std::vector<DeviceGlobalMapEntry *>
 ProgramManager::getProfileCounterDeviceGlobalEntries(
     const context_impl *CtxImpl) {
-  std::vector<DeviceGlobalMapEntry *> ProfileCounters =
+  const std::vector<DeviceGlobalMapEntry *> ProfileCounters =
       ProgramManager::getInstance().m_DeviceGlobals.getProfileCounterEntries();
   std::vector<DeviceGlobalMapEntry *> FoundEntries;
   for (const auto &DGEntry : ProfileCounters) {

sycl/test-e2e/Basic/device_code_coverage.cpp renamed to sycl/test-e2e/Coverage/device_code_coverage.cpp

@@ -1,9 +1,9 @@
 // RUN: %{build} -fprofile-instr-generate -fcoverage-mapping -o %t.out
 // RUN: %{run} LLVM_PROFILE_FILE=%t.profraw %t.out
-// RUN: llvm-profdata merge %t.profraw -o %t.profdata
-// RUN: llvm-cov show -instr-profile=%t.profdata %t.out -name="main" | FileCheck %s
+// RUN: %{run-aux} llvm-profdata merge %t.profraw -o %t.profdata
+// RUN: %{run-aux} llvm-cov show -instr-profile=%t.profdata %t.out -name="main" | FileCheck %s
 
-#include <sycl/sycl.hpp>
+#include <sycl/usm.hpp>
 
 int main() {
   sycl::queue q;
@@ -22,6 +22,7 @@ int main() {
   return 0;
 }
 
+// REQUIRES: target-spir
 // UNSUPPORTED: opencl && gpu
 // UNSUPPORTED-TRACKER: GSD-4287