[SYCL] Fix warnings in source/detail code (#16334)

Fixes places with copy instead of usage by reference or move.
Protected potential but unlikely constant overflow with exception.

---------

Signed-off-by: Tikhomirova, Kseniya <kseniya.tikhomirova@intel.com>

sycl/source/detail/helpers.cpp

@@ -94,7 +94,7 @@ retrieveKernelBinary(const QueueImplPtr &Queue, const char *KernelName,
     DeviceImage = &detail::ProgramManager::getInstance().getDeviceImage(
         KernelName, Context, Device);
     Program = detail::ProgramManager::getInstance().createURProgram(
-        *DeviceImage, Context, {Device});
+        *DeviceImage, Context, {std::move(Device)});
   }
   return {DeviceImage, Program};
 }

sycl/source/detail/persistent_device_code_cache.cpp

@@ -320,7 +320,7 @@ std::vector<std::vector<char>> PersistentDeviceCodeCache::getItemFromDisc(
 std::vector<std::vector<char>>
 PersistentDeviceCodeCache::getCompiledKernelFromDisc(
     const std::vector<device> &Devices, const std::string &BuildOptionsString,
-    const std::string SourceStr) {
+    const std::string &SourceStr) {
   assert(!Devices.empty());
   std::vector<std::vector<char>> Binaries(Devices.size());
   std::string FileNames;
@@ -518,7 +518,7 @@ std::string PersistentDeviceCodeCache::getCacheItemPath(
 
 std::string PersistentDeviceCodeCache::getCompiledKernelItemPath(
     const device &Device, const std::string &BuildOptionsString,
-    const std::string SourceString) {
+    const std::string &SourceString) {
 
   std::string cache_root{getRootDir()};
   if (cache_root.empty()) {

sycl/source/detail/persistent_device_code_cache.hpp

@@ -170,7 +170,7 @@ class PersistentDeviceCodeCache {
   static std::string
   getCompiledKernelItemPath(const device &Device,
                             const std::string &BuildOptionsString,
-                            const std::string SourceString);
+                            const std::string &SourceString);
 
   /* Program binaries built for one or more devices are read from persistent
    * cache and returned in form of vector of programs. Each binary program is
@@ -185,7 +185,7 @@ class PersistentDeviceCodeCache {
   static std::vector<std::vector<char>>
   getCompiledKernelFromDisc(const std::vector<device> &Devices,
                             const std::string &BuildOptionsString,
-                            const std::string SourceStr);
+                            const std::string &SourceStr);
 
   /* Stores build program in persistent cache
    */

sycl/source/detail/program_manager/program_manager.cpp

@@ -76,7 +76,7 @@ static ur_program_handle_t
 createBinaryProgram(const ContextImplPtr Context,
                     const std::vector<device> &Devices,
                     const uint8_t **Binaries, size_t *Lengths,
-                    const std::vector<ur_program_metadata_t> Metadata) {
+                    const std::vector<ur_program_metadata_t> &Metadata) {
   const AdapterPtr &Adapter = Context->getAdapter();
   ur_program_handle_t Program;
   std::vector<ur_device_handle_t> DeviceHandles;
@@ -230,7 +230,7 @@ ProgramManager::createURProgram(const RTDeviceBinaryImage &Img,
         "SPIR-V online compilation is not supported in this context");
 
   // Get program metadata from properties
-  auto ProgMetadata = Img.getProgramMetadataUR();
+  const auto &ProgMetadata = Img.getProgramMetadataUR();
 
   // Load the image
   const ContextImplPtr Ctx = getSyclObjImpl(Context);
@@ -990,7 +990,15 @@ ur_program_handle_t ProgramManager::getBuiltURProgram(
     // emplace all subsets of the current set of devices into the cache.
     // Set of all devices is not included in the loop as it was already added
     // into the cache.
-    for (int Mask = 1; Mask < (1 << URDevicesSet.size()) - 1; ++Mask) {
+    int Mask = 1;
+    if (URDevicesSet.size() > sizeof(Mask) * 8 - 1) {
+      // Protection for the algorithm below. Although overflow is very unlikely
+      // to be reached.
+      throw sycl::exception(
+          make_error_code(errc::runtime),
+          "Unable to cache built program for more than 31 devices");
+    }
+    for (; Mask < (1 << URDevicesSet.size()) - 1; ++Mask) {
       std::set<ur_device_handle_t> Subset;
       int Index = 0;
       for (auto It = URDevicesSet.begin(); It != URDevicesSet.end();
@@ -1124,7 +1132,7 @@ ProgramManager::getUrProgramFromUrKernel(ur_kernel_handle_t Kernel,
 
 std::string
 ProgramManager::getProgramBuildLog(const ur_program_handle_t &Program,
-                                   const ContextImplPtr Context) {
+                                   const ContextImplPtr &Context) {
   size_t URDevicesSize = 0;
   const AdapterPtr &Adapter = Context->getAdapter();
   Adapter->call<UrApiKind::urProgramGetInfo>(Program, UR_PROGRAM_INFO_DEVICES,

sycl/source/detail/program_manager/program_manager.hpp

@@ -220,7 +220,7 @@ class ProgramManager {
   void addImages(sycl_device_binaries DeviceImages);
   void debugPrintBinaryImages() const;
   static std::string getProgramBuildLog(const ur_program_handle_t &Program,
-                                        const ContextImplPtr Context);
+                                        const ContextImplPtr &Context);
 
   uint32_t getDeviceLibReqMask(const RTDeviceBinaryImage &Img);
 

sycl/source/detail/scheduler/commands.cpp

@@ -59,7 +59,7 @@ namespace detail {
 template <typename MemOpFuncT, typename... MemOpArgTs>
 ur_result_t callMemOpHelper(MemOpFuncT &MemOpFunc, MemOpArgTs &&...MemOpArgs) {
   try {
-    MemOpFunc(MemOpArgs...);
+    MemOpFunc(std::forward<MemOpArgTs>(MemOpArgs)...);
   } catch (sycl::exception &e) {
     return static_cast<ur_result_t>(get_ur_error(e));
   }
@@ -70,7 +70,7 @@ template <typename MemOpRet, typename MemOpFuncT, typename... MemOpArgTs>
 ur_result_t callMemOpHelperRet(MemOpRet &MemOpResult, MemOpFuncT &MemOpFunc,
                                MemOpArgTs &&...MemOpArgs) {
   try {
-    MemOpResult = MemOpFunc(MemOpArgs...);
+    MemOpResult = MemOpFunc(std::forward<MemOpArgTs>(MemOpArgs)...);
   } catch (sycl::exception &e) {
     return static_cast<ur_result_t>(get_ur_error(e));
   }
@@ -2891,7 +2891,7 @@ ur_result_t ExecCGCommand::enqueueImpCommandBuffer() {
                                                        &RawEvents[0]);
   }
 
-  ur_exp_command_buffer_sync_point_t OutSyncPoint;
+  ur_exp_command_buffer_sync_point_t OutSyncPoint{};
   ur_exp_command_buffer_command_handle_t OutCommand = nullptr;
   switch (MCommandGroup->getType()) {
   case CGType::Kernel: {