From b61d53447203ca7a321d72f6bdd3f553a3aa06c4 Mon Sep 17 00:00:00 2001 From: Jiaming Yuan Date: Mon, 14 Oct 2019 09:13:33 -0400 Subject: [PATCH] Span: use `size_t' for index_type, add `front' and `back'. (#4935) * Use `size_t' for index_type. Add `front' and `back'. * Remove a batch of `static_cast'. --- include/xgboost/span.h | 112 ++++++++++++++------------- src/common/device_helpers.cuh | 13 ++-- src/common/hist_util.h | 2 +- src/common/host_device_vector.cu | 5 +- src/data/columnar.h | 22 +++--- tests/cpp/common/test_span.cc | 48 +++++++++--- tests/cpp/common/test_span.cu | 43 +++++++++- tests/cpp/data/test_data.cc | 4 +- tests/cpp/tree/test_quantile_hist.cc | 4 +- 9 files changed, 160 insertions(+), 93 deletions(-) diff --git a/include/xgboost/span.h b/include/xgboost/span.h index d7e48ecb9..13a363e17 100644 --- a/include/xgboost/span.h +++ b/include/xgboost/span.h @@ -31,7 +31,8 @@ #include // CHECK -#include // int64_t +#include // size_t +#include // numeric_limits #include /*! @@ -97,18 +98,20 @@ namespace detail { * represent ptrdiff_t, which is just int64_t. So we make it determinstic * here. */ -using ptrdiff_t = int64_t; // NOLINT +using ptrdiff_t = typename std::conditional::value, + std::ptrdiff_t, std::int64_t>::type; } // namespace detail #if defined(_MSC_VER) && _MSC_VER < 1910 -constexpr const detail::ptrdiff_t dynamic_extent = -1; // NOLINT +constexpr const std::size_t +dynamic_extent = std::numeric_limits::max(); // NOLINT #else -constexpr detail::ptrdiff_t dynamic_extent = -1; // NOLINT +constexpr std::size_t dynamic_extent = std::numeric_limits::max(); // NOLINT #endif // defined(_MSC_VER) && _MSC_VER < 1910 enum class byte : unsigned char {}; // NOLINT -template +template class Span; namespace detail { @@ -119,8 +122,8 @@ class SpanIterator { public: using iterator_category = std::random_access_iterator_tag; // NOLINT - using value_type = typename std::remove_cv::type; // NOLINT - using difference_type = typename SpanType::index_type; // NOLINT + using value_type = typename SpanType::value_type; // NOLINT + using difference_type = detail::ptrdiff_t; // NOLINT using reference = typename std::conditional< // NOLINT IsConst, const ElementType, ElementType>::type&; @@ -153,7 +156,7 @@ class SpanIterator { } XGBOOST_DEVICE SpanIterator& operator++() { - SPAN_CHECK(0 <= index_ && index_ != span_->size()); + SPAN_CHECK(index_ != span_->size()); index_++; return *this; } @@ -182,7 +185,7 @@ class SpanIterator { } XGBOOST_DEVICE SpanIterator& operator+=(difference_type n) { - SPAN_CHECK((index_ + n) >= 0 && (index_ + n) <= span_->size()); + SPAN_CHECK((index_ + n) <= span_->size()); index_ += n; return *this; } @@ -234,7 +237,7 @@ class SpanIterator { protected: const SpanType *span_; - detail::ptrdiff_t index_; + typename SpanType::index_type index_; }; @@ -248,24 +251,22 @@ class SpanIterator { * - Otherwise, if Extent is not dynamic_extent, Extent - Offset; * - Otherwise, dynamic_extent. */ -template +template struct ExtentValue : public std::integral_constant< - detail::ptrdiff_t, Count != dynamic_extent ? + std::size_t, Count != dynamic_extent ? Count : (Extent != dynamic_extent ? Extent - Offset : Extent)> {}; /*! * If N is dynamic_extent, the extent of the returned span E is also - * dynamic_extent; otherwise it is detail::ptrdiff_t(sizeof(T)) * N. + * dynamic_extent; otherwise it is std::size_t(sizeof(T)) * N. */ -template +template struct ExtentAsBytesValue : public std::integral_constant< - detail::ptrdiff_t, + std::size_t, Extent == dynamic_extent ? - Extent : static_cast(sizeof(T) * Extent)> {}; + Extent : sizeof(T) * Extent> {}; -template +template struct IsAllowedExtentConversion : public std::integral_constant< bool, From == To || From == dynamic_extent || To == dynamic_extent> {}; @@ -276,7 +277,7 @@ struct IsAllowedElementTypeConversion : public std::integral_constant< template struct IsSpanOracle : std::false_type {}; -template +template struct IsSpanOracle> : std::true_type {}; template @@ -385,12 +386,12 @@ XGBOOST_DEVICE bool LexicographicalCompare(InputIt1 first1, InputIt1 last1, * passing iterator. */ template + std::size_t Extent = dynamic_extent> class Span { public: using element_type = T; // NOLINT using value_type = typename std::remove_cv::type; // NOLINT - using index_type = detail::ptrdiff_t; // NOLINT + using index_type = std::size_t; // NOLINT using difference_type = detail::ptrdiff_t; // NOLINT using pointer = T*; // NOLINT using reference = T&; // NOLINT @@ -406,13 +407,12 @@ class Span { XGBOOST_DEVICE Span(pointer _ptr, index_type _count) : size_(_count), data_(_ptr) { - SPAN_CHECK(_count >= 0); + SPAN_CHECK(!(Extent != dynamic_extent && _count != Extent)); SPAN_CHECK(_ptr || _count == 0); } XGBOOST_DEVICE Span(pointer _first, pointer _last) : size_(_last - _first), data_(_first) { - SPAN_CHECK(size_ >= 0); SPAN_CHECK(data_ || size_ == 0); } @@ -441,7 +441,7 @@ class Span { XGBOOST_DEVICE Span(const Container& _cont) : size_(_cont.size()), // NOLINT data_(_cont.data()) {} - template ::value && detail::IsAllowedExtentConversion::value>> @@ -491,8 +491,18 @@ class Span { return const_reverse_iterator{cbegin()}; } + // element access + + XGBOOST_DEVICE reference front() const { + return (*this)[0]; + } + + XGBOOST_DEVICE reference back() const { + return (*this)[size() - 1]; + } + XGBOOST_DEVICE reference operator[](index_type _idx) const { - SPAN_CHECK(_idx >= 0 && _idx < size()); + SPAN_CHECK(_idx < size()); return data()[_idx]; } @@ -517,27 +527,27 @@ class Span { } // Subviews - template + template XGBOOST_DEVICE Span first() const { // NOLINT - SPAN_CHECK(Count >= 0 && Count <= size()); + SPAN_CHECK(Count <= size()); return {data(), Count}; } XGBOOST_DEVICE Span first( // NOLINT - detail::ptrdiff_t _count) const { - SPAN_CHECK(_count >= 0 && _count <= size()); + std::size_t _count) const { + SPAN_CHECK(_count <= size()); return {data(), _count}; } - template + template XGBOOST_DEVICE Span last() const { // NOLINT - SPAN_CHECK(Count >=0 && size() - Count >= 0); + SPAN_CHECK(Count <= size()); return {data() + size() - Count, Count}; } XGBOOST_DEVICE Span last( // NOLINT - detail::ptrdiff_t _count) const { - SPAN_CHECK(_count >= 0 && _count <= size()); + std::size_t _count) const { + SPAN_CHECK(_count <= size()); return subspan(size() - _count, _count); } @@ -545,24 +555,22 @@ class Span { * If Count is std::dynamic_extent, r.size() == this->size() - Offset; * Otherwise r.size() == Count. */ - template + template XGBOOST_DEVICE auto subspan() const -> // NOLINT Span::value> { - SPAN_CHECK(Offset >= 0 && (Offset < size() || size() == 0)); - SPAN_CHECK(Count == dynamic_extent || - (Count >= 0 && Offset + Count <= size())); + SPAN_CHECK(Offset < size() || size() == 0); + SPAN_CHECK(Count == dynamic_extent || (Offset + Count <= size())); return {data() + Offset, Count == dynamic_extent ? size() - Offset : Count}; } XGBOOST_DEVICE Span subspan( // NOLINT - detail::ptrdiff_t _offset, - detail::ptrdiff_t _count = dynamic_extent) const { - SPAN_CHECK(_offset >= 0 && (_offset < size() || size() == 0)); - SPAN_CHECK((_count == dynamic_extent) || - (_count >= 0 && _offset + _count <= size())); + index_type _offset, + index_type _count = dynamic_extent) const { + SPAN_CHECK(_offset < size() || size() == 0); + SPAN_CHECK((_count == dynamic_extent) || (_offset + _count <= size())); return {data() + _offset, _count == dynamic_extent ? size() - _offset : _count}; @@ -573,7 +581,7 @@ class Span { pointer data_; }; -template +template XGBOOST_DEVICE bool operator==(Span l, Span r) { if (l.size() != r.size()) { return false; @@ -587,23 +595,23 @@ XGBOOST_DEVICE bool operator==(Span l, Span r) { return true; } -template +template XGBOOST_DEVICE constexpr bool operator!=(Span l, Span r) { return !(l == r); } -template +template XGBOOST_DEVICE constexpr bool operator<(Span l, Span r) { return detail::LexicographicalCompare(l.begin(), l.end(), r.begin(), r.end()); } -template +template XGBOOST_DEVICE constexpr bool operator<=(Span l, Span r) { return !(l > r); } -template +template XGBOOST_DEVICE constexpr bool operator>(Span l, Span r) { return detail::LexicographicalCompare< typename Span::iterator, typename Span::iterator, @@ -611,18 +619,18 @@ XGBOOST_DEVICE constexpr bool operator>(Span l, Span r) { r.begin(), r.end()); } -template +template XGBOOST_DEVICE constexpr bool operator>=(Span l, Span r) { return !(l < r); } -template +template XGBOOST_DEVICE auto as_bytes(Span s) __span_noexcept -> // NOLINT Span::value> { return {reinterpret_cast(s.data()), s.size_bytes()}; } -template +template XGBOOST_DEVICE auto as_writable_bytes(Span s) __span_noexcept -> // NOLINT Span::value> { return {reinterpret_cast(s.data()), s.size_bytes()}; diff --git a/src/common/device_helpers.cuh b/src/common/device_helpers.cuh index ede28cd42..533a62954 100644 --- a/src/common/device_helpers.cuh +++ b/src/common/device_helpers.cuh @@ -380,9 +380,7 @@ class DoubleBuffer { T *Current() { return buff.Current(); } xgboost::common::Span CurrentSpan() { - return xgboost::common::Span{ - buff.Current(), - static_cast::index_type>(Size())}; + return xgboost::common::Span{buff.Current(), Size()}; } T *other() { return buff.Alternate(); } @@ -1120,17 +1118,16 @@ template ToSpan( device_vector& vec, IndexT offset = 0, - IndexT size = -1) { - size = size == -1 ? vec.size() : size; + IndexT size = std::numeric_limits::max()) { + size = size == std::numeric_limits::max() ? vec.size() : size; CHECK_LE(offset + size, vec.size()); - return {vec.data().get() + offset, static_cast(size)}; + return {vec.data().get() + offset, size}; } template xgboost::common::Span ToSpan(thrust::device_vector& vec, size_t offset, size_t size) { - using IndexT = typename xgboost::common::Span::index_type; - return ToSpan(vec, static_cast(offset), static_cast(size)); + return ToSpan(vec, offset, size); } // thrust begin, similiar to std::begin diff --git a/src/common/hist_util.h b/src/common/hist_util.h index 1ae9389d4..176f1b495 100644 --- a/src/common/hist_util.h +++ b/src/common/hist_util.h @@ -343,7 +343,7 @@ struct GHistIndexBlock { // get i-th row inline GHistIndexRow operator[](size_t i) const { - return {&index[0] + row_ptr[i], detail::ptrdiff_t(row_ptr[i + 1] - row_ptr[i])}; + return {&index[0] + row_ptr[i], row_ptr[i + 1] - row_ptr[i]}; } }; diff --git a/src/common/host_device_vector.cu b/src/common/host_device_vector.cu index 7f8f9e6b3..a85194aed 100644 --- a/src/common/host_device_vector.cu +++ b/src/common/host_device_vector.cu @@ -69,13 +69,12 @@ class HostDeviceVectorImpl { common::Span DeviceSpan() { LazySyncDevice(GPUAccess::kWrite); - return {data_d_.data().get(), static_cast::index_type>(Size())}; + return {data_d_.data().get(), Size()}; } common::Span ConstDeviceSpan() { LazySyncDevice(GPUAccess::kRead); - using SpanInd = typename common::Span::index_type; - return {data_d_.data().get(), static_cast(Size())}; + return {data_d_.data().get(), Size()}; } void Fill(T v) { // NOLINT diff --git a/src/data/columnar.h b/src/data/columnar.h index 652fd207c..b47324377 100644 --- a/src/data/columnar.h +++ b/src/data/columnar.h @@ -165,6 +165,14 @@ class ArrayInterfaceHandler { auto typestr = get(j_mask.at("typestr")); // For now this is just 1, we can support different size of interger in mask. int64_t const type_length = typestr.at(2) - 48; + + if (typestr.at(1) == 't') { + CHECK_EQ(type_length, 1) << "mask with bitfield type should be of 1 byte per bitfield."; + } else if (typestr.at(1) == 'i') { + CHECK_EQ(type_length, 1) << "mask with integer type should be of 1 byte per integer."; + } else { + LOG(FATAL) << "mask must be of integer type or bit field type."; + } /* * shape represents how many bits is in the mask. (This is a grey area, don't be * suprised if it suddently represents something else when supporting a new @@ -175,10 +183,10 @@ class ArrayInterfaceHandler { * * And that's the only requirement. */ - int64_t const n_bits = get(j_shape.at(0)); + size_t const n_bits = static_cast(get(j_shape.at(0))); // The size of span required to cover all bits. Here with 8 bits bitfield, we // assume 1 byte alignment. - int64_t const span_size = RBitField8::ComputeStorageSize(n_bits); + size_t const span_size = RBitField8::ComputeStorageSize(n_bits); if (j_mask.find("strides") != j_mask.cend()) { auto strides = get(column.at("strides")); @@ -186,14 +194,6 @@ class ArrayInterfaceHandler { CHECK_EQ(get(strides.at(0)), type_length) << ColumnarErrors::Contigious(); } - if (typestr.at(1) == 't') { - CHECK_EQ(typestr.at(2), '1') << "mask with bitfield type should be of 1 byte per bitfield."; - } else if (typestr.at(1) == 'i') { - CHECK_EQ(typestr.at(2), '1') << "mask with integer type should be of 1 byte per integer."; - } else { - LOG(FATAL) << "mask must be of integer type or bit field type."; - } - s_mask = {p_mask, span_size}; return n_bits; } @@ -219,7 +219,7 @@ class ArrayInterfaceHandler { CHECK_EQ(get(strides.at(0)), sizeof(T)) << ColumnarErrors::Contigious(); } - auto length = get(j_shape.at(0)); + auto length = static_cast(get(j_shape.at(0))); T* p_data = ArrayInterfaceHandler::GetPtrFromArrayData(column); return common::Span{p_data, length}; diff --git a/tests/cpp/common/test_span.cc b/tests/cpp/common/test_span.cc index 288cd1203..c27075b93 100644 --- a/tests/cpp/common/test_span.cc +++ b/tests/cpp/common/test_span.cc @@ -98,7 +98,8 @@ TEST(Span, FromPtrLen) { } { - EXPECT_ANY_THROW(Span tmp (arr, -1);); + auto lazy = [=]() {Span tmp (arr, 5);}; + EXPECT_ANY_THROW(lazy()); } // dynamic extent @@ -298,6 +299,32 @@ TEST(Span, Obversers) { ASSERT_EQ(status, 1); } +TEST(Span, FrontBack) { + { + float arr[4] {0, 1, 2, 3}; + Span s(arr); + ASSERT_EQ(s.front(), 0); + ASSERT_EQ(s.back(), 3); + } + { + std::vector arr {0, 1, 2, 3}; + Span s(arr); + ASSERT_EQ(s.front(), 0); + ASSERT_EQ(s.back(), 3); + } + + { + Span s; + EXPECT_ANY_THROW(s.front()); + EXPECT_ANY_THROW(s.back()); + } + { + Span s; + EXPECT_ANY_THROW(s.front()); + EXPECT_ANY_THROW(s.back()); + } +} + TEST(Span, FirstLast) { // static extent { @@ -310,11 +337,11 @@ TEST(Span, FirstLast) { ASSERT_EQ(first.size(), 4); ASSERT_EQ(first.data(), arr); - for (int64_t i = 0; i < first.size(); ++i) { + for (size_t i = 0; i < first.size(); ++i) { ASSERT_EQ(first[i], arr[i]); } - - EXPECT_ANY_THROW(s.first<-1>()); + auto constexpr kOne = static_cast::index_type>(-1); + EXPECT_ANY_THROW(s.first()); EXPECT_ANY_THROW(s.first<17>()); EXPECT_ANY_THROW(s.first<32>()); } @@ -329,11 +356,11 @@ TEST(Span, FirstLast) { ASSERT_EQ(last.size(), 4); ASSERT_EQ(last.data(), arr + 12); - for (int64_t i = 0; i < last.size(); ++i) { + for (size_t i = 0; i < last.size(); ++i) { ASSERT_EQ(last[i], arr[i+12]); } - - EXPECT_ANY_THROW(s.last<-1>()); + auto constexpr kOne = static_cast::index_type>(-1); + EXPECT_ANY_THROW(s.last()); EXPECT_ANY_THROW(s.last<17>()); EXPECT_ANY_THROW(s.last<32>()); } @@ -348,7 +375,7 @@ TEST(Span, FirstLast) { ASSERT_EQ(first.size(), 4); ASSERT_EQ(first.data(), s.data()); - for (int64_t i = 0; i < first.size(); ++i) { + for (size_t i = 0; i < first.size(); ++i) { ASSERT_EQ(first[i], s[i]); } @@ -368,7 +395,7 @@ TEST(Span, FirstLast) { ASSERT_EQ(last.size(), 4); ASSERT_EQ(last.data(), s.data() + 12); - for (int64_t i = 0; i < last.size(); ++i) { + for (size_t i = 0; i < last.size(); ++i) { ASSERT_EQ(s[12 + i], last[i]); } @@ -397,7 +424,8 @@ TEST(Span, Subspan) { EXPECT_ANY_THROW(s1.subspan(-1, 0)); EXPECT_ANY_THROW(s1.subspan(16, 0)); - EXPECT_ANY_THROW(s1.subspan<-1>()); + auto constexpr kOne = static_cast::index_type>(-1); + EXPECT_ANY_THROW(s1.subspan()); EXPECT_ANY_THROW(s1.subspan<16>()); } diff --git a/tests/cpp/common/test_span.cu b/tests/cpp/common/test_span.cu index e7809b35a..9aa0b8d53 100644 --- a/tests/cpp/common/test_span.cu +++ b/tests/cpp/common/test_span.cu @@ -240,16 +240,16 @@ TEST(GPUSpan, ElementAccess) { } __global__ void TestFirstDynamicKernel(Span _span) { - _span.first<-1>(); + _span.first::index_type>(-1)>(); } __global__ void TestFirstStaticKernel(Span _span) { - _span.first(-1); + _span.first(static_cast::index_type>(-1)); } __global__ void TestLastDynamicKernel(Span _span) { - _span.last<-1>(); + _span.last::index_type>(-1)>(); } __global__ void TestLastStaticKernel(Span _span) { - _span.last(-1); + _span.last(static_cast::index_type>(-1)); } TEST(GPUSpan, FirstLast) { @@ -312,6 +312,41 @@ TEST(GPUSpan, FirstLast) { output = testing::internal::GetCapturedStdout(); } +__global__ void TestFrontKernel(Span _span) { + _span.front(); +} + +__global__ void TestBackKernel(Span _span) { + _span.back(); +} + +TEST(GPUSpan, FrontBack) { + dh::safe_cuda(cudaSetDevice(0)); + + Span s; + auto lambda_test_front = [=]() { + // make sure the termination happens inside this test. + try { + TestFrontKernel<<<1, 1>>>(s); + dh::safe_cuda(cudaDeviceSynchronize()); + dh::safe_cuda(cudaGetLastError()); + } catch (dmlc::Error const& e) { + std::terminate(); + } + }; + EXPECT_DEATH(lambda_test_front(), ""); + + auto lambda_test_back = [=]() { + try { + TestBackKernel<<<1, 1>>>(s); + dh::safe_cuda(cudaDeviceSynchronize()); + dh::safe_cuda(cudaGetLastError()); + } catch (dmlc::Error const& e) { + std::terminate(); + } + }; + EXPECT_DEATH(lambda_test_back(), ""); +} __global__ void TestSubspanDynamicKernel(Span _span) { _span.subspan(16, 0); diff --git a/tests/cpp/data/test_data.cc b/tests/cpp/data/test_data.cc index 173330e56..2f9926857 100644 --- a/tests/cpp/data/test_data.cc +++ b/tests/cpp/data/test_data.cc @@ -50,7 +50,7 @@ TEST(SparsePage, PushCSC) { inst = page[1]; ASSERT_EQ(inst.size(), 6); std::vector indices_sol {1, 2, 3}; - for (int64_t i = 0; i < inst.size(); ++i) { + for (size_t i = 0; i < inst.size(); ++i) { ASSERT_EQ(inst[i].index, indices_sol[i % 3]); } } @@ -76,7 +76,7 @@ TEST(SparsePage, PushCSCAfterTranspose) { // how the dmatrix has been created for (size_t i = 0; i < page.Size(); ++i) { auto inst = page[i]; - for (int j = 1; j < inst.size(); ++j) { + for (size_t j = 1; j < inst.size(); ++j) { ASSERT_EQ(inst[0].fvalue, inst[j].fvalue); } } diff --git a/tests/cpp/tree/test_quantile_hist.cc b/tests/cpp/tree/test_quantile_hist.cc index 0a999953b..2561a1383 100644 --- a/tests/cpp/tree/test_quantile_hist.cc +++ b/tests/cpp/tree/test_quantile_hist.cc @@ -73,7 +73,7 @@ class QuantileHistMock : public QuantileHistMaker { ASSERT_LT(gmat_row_offset, gmat.index.size()); SparsePage::Inst inst = batch[i]; ASSERT_EQ(gmat.row_ptr[rid] + inst.size(), gmat.row_ptr[rid + 1]); - for (int64_t j = 0; j < inst.size(); ++j) { + for (size_t j = 0; j < inst.size(); ++j) { // Each entry of GHistIndexMatrix represents a bin ID const size_t bin_id = gmat.index[gmat_row_offset + j]; const size_t fid = inst[j].index; @@ -129,7 +129,7 @@ class QuantileHistMock : public QuantileHistMaker { } // Now validate the computed histogram returned by BuildHist - for (int64_t i = 0; i < hist_[nid].size(); ++i) { + for (size_t i = 0; i < hist_[nid].size(); ++i) { GradientPairPrecise sol = histogram_expected[i]; ASSERT_NEAR(sol.GetGrad(), hist_[nid][i].GetGrad(), kEps); ASSERT_NEAR(sol.GetHess(), hist_[nid][i].GetHess(), kEps);