Hendrik/xgboost
1
0
Fork 0
Code Issues Pull Requests Packages Projects Releases Wiki Activity

Dask device dmatrix (#5901)

Browse Source 
* Fix softprob with empty dmatrix.
This commit is contained in:
Jiaming Yuan
2020-07-17 13:17:43 +08:00
committed by GitHub
parent e471056ec4
commit 7c2686146e
12 changed files with 392 additions and 149 deletions
Download Patch File Download Diff File Expand all files Collapse all files

1
src/common/quantile.cuh
View File

@@ -85,6 +85,7 @@ class SketchContainer {
// Initialize Sketches for this dmatrix
this->columns_ptr_.SetDevice(device_);
this->columns_ptr_.Resize(num_columns + 1);
CHECK_GE(device, 0);
timer_.Init(__func__);
}
/* \brief Return GPU ID for this container. */

5
src/data/array_interface.h
View File

@@ -114,7 +114,6 @@ class ArrayInterfaceHandler {
get<Array const>(
obj.at("data"))
.at(0))));
CHECK(p_data);
return p_data;
}
@@ -224,6 +223,9 @@ class ArrayInterfaceHandler {
auto shape = ExtractShape(column);
T* p_data = ArrayInterfaceHandler::GetPtrFromArrayData<T*>(column);
if (!p_data) {
CHECK_EQ(shape.first * shape.second, 0) << "Empty data with non-zero shape.";
}
return common::Span<T>{p_data, shape.first * shape.second};
}
};
@@ -234,7 +236,6 @@ class ArrayInterface {
bool allow_mask = true) {
ArrayInterfaceHandler::Validate(column);
data = ArrayInterfaceHandler::GetPtrFromArrayData<void*>(column);
CHECK(data) << "Column is null";
auto shape = ArrayInterfaceHandler::ExtractShape(column);
num_rows = shape.first;
num_cols = shape.second;

9
src/data/data.cc
View File

@@ -488,6 +488,15 @@ void MetaInfo::Extend(MetaInfo const& that, bool accumulate_rows) {
this->group_ptr_.insert(this->group_ptr_.end(), group_ptr.begin() + 1,
group_ptr.end());
}
if (!that.feature_names.empty()) {
this->feature_names = that.feature_names;
}
if (!that.feature_type_names.empty()) {
this->feature_type_names = that.feature_type_names;
auto &h_feature_types = feature_types.HostVector();
LoadFeatureType(this->feature_type_names, &h_feature_types);
}
}
void MetaInfo::Validate(int32_t device) const {

3
src/data/data.cu
View File

@@ -69,6 +69,9 @@ void MetaInfo::SetInfo(const char * c_key, std::string const& interface_str) {
<< "Meta info " << key << " should be dense, found validity mask";
CHECK_EQ(array_interface.num_cols, 1)
<< "Meta info should be a single column.";
if (array_interface.num_rows == 0) {
return;
}
if (key == "label") {
CopyInfoImpl(array_interface, &labels_);

11
src/data/device_adapter.cuh
View File

@@ -122,10 +122,14 @@ class CudfAdapter : public detail::SingleBatchDataIter<CudfAdapterBatch> {
CHECK_NE(typestr.front(), '>') << ArrayInterfaceErrors::BigEndian();
std::vector<ArrayInterface> columns;
auto first_column = ArrayInterface(get<Object const>(json_columns[0]));
num_rows_ = first_column.num_rows;
if (num_rows_ == 0) {
return;
}
device_idx_ = dh::CudaGetPointerDevice(first_column.data);
CHECK_NE(device_idx_, -1);
dh::safe_cuda(cudaSetDevice(device_idx_));
num_rows_ = first_column.num_rows;
for (auto& json_col : json_columns) {
auto column = ArrayInterface(get<Object const>(json_col));
columns.push_back(column);
@@ -183,9 +187,12 @@ class CupyAdapter : public detail::SingleBatchDataIter<CupyAdapterBatch> {
Json json_array_interface =
Json::Load({cuda_interface_str.c_str(), cuda_interface_str.size()});
array_interface_ = ArrayInterface(get<Object const>(json_array_interface), false);
batch_ = CupyAdapterBatch(array_interface_);
if (array_interface_.num_rows == 0) {
return;
}
device_idx_ = dh::CudaGetPointerDevice(array_interface_.data);
CHECK_NE(device_idx_, -1);
batch_ = CupyAdapterBatch(array_interface_);
}
const CupyAdapterBatch& Value() const override { return batch_; }

48
src/data/iterative_device_dmatrix.cu
View File

@@ -62,23 +62,30 @@ void IterativeDeviceDMatrix::Initialize(DataIterHandle iter_handle, float missin
size_t batches = 0;
size_t accumulated_rows = 0;
bst_feature_t cols = 0;
int32_t device = -1;
int32_t device = GenericParameter::kCpuId;
int32_t current_device_;
dh::safe_cuda(cudaGetDevice(&current_device_));
auto get_device = [&]() -> int32_t {
int32_t d = GenericParameter::kCpuId ? current_device_ : device;
return d;
};
while (iter.Next()) {
device = proxy->DeviceIdx();
dh::safe_cuda(cudaSetDevice(device));
dh::safe_cuda(cudaSetDevice(get_device()));
if (cols == 0) {
cols = num_cols();
rabit::Allreduce<rabit::op::Max>(&cols, 1);
} else {
CHECK_EQ(cols, num_cols()) << "Inconsistent number of columns.";
}
sketch_containers.emplace_back(batch_param_.max_bin, num_cols(), num_rows(), device);
sketch_containers.emplace_back(batch_param_.max_bin, cols, num_rows(), get_device());
auto* p_sketch = &sketch_containers.back();
proxy->Info().weights_.SetDevice(device);
proxy->Info().weights_.SetDevice(get_device());
Dispatch(proxy, [&](auto const &value) {
common::AdapterDeviceSketch(value, batch_param_.max_bin,
proxy->Info(), missing, p_sketch);
});
auto batch_rows = num_rows();
accumulated_rows += batch_rows;
dh::caching_device_vector<size_t> row_counts(batch_rows + 1, 0);
@@ -86,19 +93,15 @@ void IterativeDeviceDMatrix::Initialize(DataIterHandle iter_handle, float missin
row_counts.size());
row_stride = std::max(row_stride, Dispatch(proxy, [=](auto const &value) {
return GetRowCounts(value, row_counts_span,
device, missing);
get_device(), missing);
}));
nnz += thrust::reduce(thrust::cuda::par(alloc), row_counts.begin(),
row_counts.end());
batches++;
}
if (device < 0) { // error or empty
this->page_.reset(new EllpackPage);
return;
}
common::SketchContainer final_sketch(batch_param_.max_bin, cols, accumulated_rows, device);
iter.Reset();
dh::safe_cuda(cudaSetDevice(get_device()));
common::SketchContainer final_sketch(batch_param_.max_bin, cols, accumulated_rows, get_device());
for (auto const& sketch : sketch_containers) {
final_sketch.Merge(sketch.ColumnsPtr(), sketch.Data());
final_sketch.FixError();
@@ -113,14 +116,14 @@ void IterativeDeviceDMatrix::Initialize(DataIterHandle iter_handle, float missin
this->info_.num_row_ = accumulated_rows;
this->info_.num_nonzero_ = nnz;
auto init_page = [this, &proxy, &cuts, row_stride, accumulated_rows]() {
auto init_page = [this, &proxy, &cuts, row_stride, accumulated_rows,
get_device]() {
if (!page_) {
// Should be put inside the while loop to protect against empty batch. In
// that case device id is invalid.
page_.reset(new EllpackPage);
*(page_->Impl()) =
EllpackPageImpl(proxy->DeviceIdx(), cuts, this->IsDense(), row_stride,
accumulated_rows);
*(page_->Impl()) = EllpackPageImpl(get_device(), cuts, this->IsDense(),
row_stride, accumulated_rows);
}
};
@@ -130,21 +133,20 @@ void IterativeDeviceDMatrix::Initialize(DataIterHandle iter_handle, float missin
size_t n_batches_for_verification = 0;
while (iter.Next()) {
init_page();
auto device = proxy->DeviceIdx();
dh::safe_cuda(cudaSetDevice(device));
dh::safe_cuda(cudaSetDevice(get_device()));
auto rows = num_rows();
dh::caching_device_vector<size_t> row_counts(rows + 1, 0);
common::Span<size_t> row_counts_span(row_counts.data().get(),
row_counts.size());
Dispatch(proxy, [=](auto const& value) {
return GetRowCounts(value, row_counts_span, device, missing);
return GetRowCounts(value, row_counts_span, get_device(), missing);
});
auto is_dense = this->IsDense();
auto new_impl = Dispatch(proxy, [&](auto const &value) {
return EllpackPageImpl(value, missing, device, is_dense, nthread,
row_counts_span, row_stride, rows, cols, cuts);
return EllpackPageImpl(value, missing, get_device(), is_dense, nthread,
row_counts_span, row_stride, rows, cols, cuts);
});
size_t num_elements = page_->Impl()->Copy(device, &new_impl, offset);
size_t num_elements = page_->Impl()->Copy(get_device(), &new_impl, offset);
offset += num_elements;
proxy->Info().num_row_ = num_rows();

28
src/metric/multiclass_metric.cu
View File

@@ -158,18 +158,22 @@ struct EvalMClassBase : public Metric {
bst_float Eval(const HostDeviceVector<bst_float> &preds,
const MetaInfo &info,
bool distributed) override {
CHECK_NE(info.labels_.Size(), 0U) << "label set cannot be empty";
CHECK(preds.Size() % info.labels_.Size() == 0)
<< "label and prediction size not match";
const size_t nclass = preds.Size() / info.labels_.Size();
CHECK_GE(nclass, 1U)
<< "mlogloss and merror are only used for multi-class classification,"
<< " use logloss for binary classification";
int device = tparam_->gpu_id;
auto result = reducer_.Reduce(*tparam_, device, nclass, info.weights_, info.labels_, preds);
double dat[2] { result.Residue(), result.Weights() };
if (info.labels_.Size() == 0) {
CHECK_EQ(preds.Size(), 0);
} else {
CHECK(preds.Size() % info.labels_.Size() == 0) << "label and prediction size not match";
}
double dat[2] { 0.0, 0.0 };
if (info.labels_.Size() != 0) {
const size_t nclass = preds.Size() / info.labels_.Size();
CHECK_GE(nclass, 1U)
<< "mlogloss and merror are only used for multi-class classification,"
<< " use logloss for binary classification";
int device = tparam_->gpu_id;
auto result = reducer_.Reduce(*tparam_, device, nclass, info.weights_, info.labels_, preds);
dat[0] = result.Residue();
dat[1] = result.Weights();
}
if (distributed) {
rabit::Allreduce<rabit::op::Sum>(dat, 2);
}

4
src/objective/multiclass_obj.cu
View File

@@ -49,7 +49,9 @@ class SoftmaxMultiClassObj : public ObjFunction {
const MetaInfo& info,
int iter,
HostDeviceVector<GradientPair>* out_gpair) override {
CHECK_NE(info.labels_.Size(), 0U) << "label set cannot be empty";
if (info.labels_.Size() == 0) {
return;
}
CHECK(preds.Size() == (static_cast<size_t>(param_.num_class) * info.labels_.Size()))
<< "SoftmaxMultiClassObj: label size and pred size does not match.\n"
<< "label.Size() * num_class: "