fix code-review comments

Signed-off-by: Tikhomirova, Kseniya <[email protected]>

Author: KseniyaTikhomirova

sycl/source/detail/program_manager/program_manager.cpp

@@ -825,24 +825,6 @@ ur_program_handle_t ProgramManager::getBuiltURProgram(
   return getBuiltURProgram(AllImages, Context, {Device});
 }
 
-template <typename Func>
-void callFuncForAllSubsets(Func &FuncToCall,
-                           const std::set<ur_device_handle_t> &DeviceSet,
-                           std::set<ur_device_handle_t> &Subset, int index) {
-  // Add the current subset to the result list
-  if (Subset.size() && Subset.size() != DeviceSet.size()) {
-    FuncToCall(Subset);
-  }
-
-  auto it = DeviceSet.begin();
-  std::advance(it, index);
-  for (size_t i = index; i < DeviceSet.size(); i++, it++) {
-    auto InsertedEntry = Subset.insert(Subset.end(), *it);
-    callFuncForAllSubsets(FuncToCall, DeviceSet, Subset, i + 1);
-    Subset.erase(InsertedEntry);
-  }
-}
-
 ur_program_handle_t ProgramManager::getBuiltURProgram(
     const BinImgWithDeps &ImgWithDeps, const context &Context,
     const std::vector<device> &Devs, const DevImgPlainWithDeps *DevImgWithDeps,
@@ -1008,23 +990,34 @@ 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.
-    auto ExecuteForAllSubsets =
-        [&CacheKey, &Cache, &Adapter, &ResProgram,
-         &CacheLinkedImages](std::set<ur_device_handle_t> &Subset) {
-          // Change device in the cache key to reduce copying of spec const
-          // data.
-          CacheKey.second = Subset;
-          bool DidInsert = Cache.insertBuiltProgram(CacheKey, ResProgram);
-          if (DidInsert) {
-            // For every cached copy of the program, we need to increment its
-            // refcount
-            Adapter->call<UrApiKind::urProgramRetain>(ResProgram);
-          }
-          CacheLinkedImages();
-        };
-    std::set<ur_device_handle_t> Subset;
-    int Index = 0;
-    callFuncForAllSubsets(ExecuteForAllSubsets, URDevicesSet, Subset, Index);
+    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 generate device sets");
+    }
+    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();
+           ++It, ++Index) {
+        if (Mask & (1 << Index)) {
+          Subset.insert(*It);
+        }
+      }
+      // Change device in the cache key to reduce copying of spec const data.
+      CacheKey.second = Subset;
+      bool DidInsert = Cache.insertBuiltProgram(CacheKey, ResProgram);
+      if (DidInsert) {
+        // For every cached copy of the program, we need to increment its
+        // refcount
+        Adapter->call<UrApiKind::urProgramRetain>(ResProgram);
+      }
+      CacheLinkedImages();
+      // getOrBuild is not supposed to return nullptr
+      assert(BuildResult != nullptr && "Invalid build result");
+    }
   }
 
   // If caching is enabled, one copy of the program handle will be