Fix sampling (#54)

* fix samling; minor refactoring

* fix

---------

Co-authored-by: Dmitry Razdoburdin <>

Author: razdoburdin

plugin/sycl/predictor/predictor.cc

@@ -2,10 +2,6 @@
  * Copyright by Contributors 2017-2024
  */
 #include <dmlc/timer.h>
-// #pragma GCC diagnostic push
-// #pragma GCC diagnostic ignored "-Wtautological-constant-compare"
-// #pragma GCC diagnostic ignored "-W#pragma-messages"
-// #pragma GCC diagnostic pop
 
 #include <cstddef>
 #include <limits>

plugin/sycl/tree/hist_updater.cc

@@ -336,7 +336,6 @@ void HistUpdater<GradientSumT>::ExpandWithLossGuide(
 
 template <typename GradientSumT>
 void HistUpdater<GradientSumT>::Update(
-    Context const * ctx,
     xgboost::tree::TrainParam const *param,
     const common::GHistIndexMatrix &gmat,
     linalg::Matrix<GradientPair> *gpair,
@@ -485,6 +484,10 @@ void HistUpdater<GradientSumT>::InitData(
   builder_monitor_.Start("InitData");
   const auto& info = fmat.Info();
 
+  if (!column_sampler_) {
+    column_sampler_ = xgboost::common::MakeColumnSampler(ctx_);
+  }
+
   // initialize the row set
   {
     row_set_collection_.Clear();
@@ -575,9 +578,9 @@ void HistUpdater<GradientSumT>::InitData(
 
   // store a pointer to the tree
   p_last_tree_ = &tree;
-  column_sampler_.Init(ctx_, info.num_col_, info.feature_weights.ConstHostVector(),
-                       param_.colsample_bynode, param_.colsample_bylevel,
-                       param_.colsample_bytree);
+  column_sampler_->Init(ctx_, info.num_col_, info.feature_weights.ConstHostVector(),
+                        param_.colsample_bynode, param_.colsample_bylevel,
+                        param_.colsample_bytree);
   if (data_layout_ == kDenseDataZeroBased || data_layout_ == kDenseDataOneBased) {
     /* specialized code for dense data:
        choose the column that has a least positive number of discrete bins.
@@ -597,8 +600,10 @@ void HistUpdater<GradientSumT>::InitData(
     }
     CHECK_GT(min_nbins_per_feature, 0U);
   }
-
-  std::fill(snode_host_.begin(), snode_host_.end(),  NodeEntry<GradientSumT>(param_));
+  {
+    qu_.wait_and_throw();
+    std::fill(snode_host_.begin(), snode_host_.end(),  NodeEntry<GradientSumT>(param_));
+  }
 
   {
     if (param_.grow_policy == xgboost::tree::TrainParam::kLossGuide) {
@@ -628,7 +633,7 @@ void HistUpdater<GradientSumT>::EvaluateSplits(
   size_t total_features = 0;
   for (size_t nid_in_set = 0; nid_in_set < n_nodes_in_set; ++nid_in_set) {
     const int32_t nid = nodes_set[nid_in_set].nid;
-    features_sets[nid_in_set] = column_sampler_.GetFeatureSet(tree.GetDepth(nid));
+    features_sets[nid_in_set] = column_sampler_->GetFeatureSet(tree.GetDepth(nid));
     for (size_t idx = 0; idx < features_sets[nid_in_set]->Size(); idx++) {
       const auto fid = features_sets[nid_in_set]->ConstHostVector()[idx];
       if (interaction_constraints_.Query(nid, fid)) {

plugin/sycl/tree/hist_updater.h

@@ -62,18 +62,19 @@ class HistUpdater {
       tree_evaluator_(qu, param, fmat->Info().num_col_),
       pruner_(std::move(pruner)),
       interaction_constraints_{std::move(int_constraints_)},
-      p_last_tree_(nullptr), p_last_fmat_(fmat),
-      column_sampler_(seed_) {
+      p_last_tree_(nullptr), p_last_fmat_(fmat) {
     builder_monitor_.Init("SYCL::Quantile::HistUpdater");
     kernel_monitor_.Init("SYCL::Quantile::HistUpdater");
+    if (param.max_depth > 0) {
+      snode_device_.Resize(&qu, 1u << (param.max_depth + 1));
+    }
     const auto sub_group_sizes =
       qu_.get_device().get_info<::sycl::info::device::sub_group_sizes>();
     sub_group_size_ = sub_group_sizes.back();
   }
 
   // update one tree, growing
-  void Update(Context const * ctx,
-              xgboost::tree::TrainParam const *param,
+  void Update(xgboost::tree::TrainParam const *param,
               const common::GHistIndexMatrix &gmat,
               linalg::Matrix<GradientPair> *gpair,
               const USMVector<GradientPair, MemoryType::on_device>& gpair_device,
@@ -181,11 +182,6 @@ class HistUpdater {
                       RegTree *p_tree,
                       const USMVector<GradientPair, MemoryType::on_device> &gpair);
 
-  void ExpandWithLossGuide(const common::GHistIndexMatrix& gmat,
-                           DMatrix* p_fmat,
-                           RegTree* p_tree,
-                           const USMVector<GradientPair, MemoryType::on_device>& gpair);
-
   void EvaluateAndApplySplits(const common::GHistIndexMatrix &gmat,
                               RegTree *p_tree,
                               int *num_leaves,
@@ -200,6 +196,11 @@ class HistUpdater {
             std::vector<ExpandEntry>* nodes_for_apply_split,
             std::vector<ExpandEntry>* temp_qexpand_depth);
 
+  void ExpandWithLossGuide(const common::GHistIndexMatrix& gmat,
+                           DMatrix* p_fmat,
+                           RegTree* p_tree,
+                           const USMVector<GradientPair, MemoryType::on_device>& gpair);
+
   void ReduceHists(const std::vector<int>& sync_ids, size_t nbins);
 
   inline static bool LossGuide(ExpandEntry lhs, ExpandEntry rhs) {
@@ -213,13 +214,14 @@ class HistUpdater {
   //  --data fields--
   const Context* ctx_;
   size_t sub_group_size_;
+  const xgboost::tree::TrainParam& param_;
+  std::shared_ptr<xgboost::common::ColumnSampler> column_sampler_;
 
   // the internal row sets
   common::RowSetCollection row_set_collection_;
   std::vector<SplitQuery> split_queries_host_;
   USMVector<SplitQuery, MemoryType::on_device> split_queries_device_;
 
-  const xgboost::tree::TrainParam& param_;
   TreeEvaluator<GradientSumT> tree_evaluator_;
   std::unique_ptr<TreeUpdater> pruner_;
   FeatureInteractionConstraintHost interaction_constraints_;
@@ -258,7 +260,6 @@ class HistUpdater {
   uint32_t fid_least_bins_;
 
   uint64_t seed_ = 0;
-  xgboost::common::ColumnSampler column_sampler_;
 
   common::PartitionBuilder partition_builder_;
 

plugin/sycl/tree/updater_quantile_hist.cc

@@ -80,7 +80,7 @@ void QuantileHistMaker::CallUpdate(
   qu_.wait();
 
   for (auto tree : trees) {
-    pimpl->Update(ctx_, param, gmat_, gpair, gpair_device_, dmat, out_position, tree);
+    pimpl->Update(param, gmat_, gpair, gpair_device_, dmat, out_position, tree);
   }
 }