From 58b0185d3458d0b495d5e2917de78bf94ead9304 Mon Sep 17 00:00:00 2001 From: Wonchan Lee Date: Tue, 23 Feb 2021 13:55:56 -0800 Subject: [PATCH 1/9] Add a failing test case that passes a leading null to scan with skipna=False --- cpp/tests/reductions/scan_tests.cpp | 26 ++++++++++++++++++++++++++ 1 file changed, 26 insertions(+) diff --git a/cpp/tests/reductions/scan_tests.cpp b/cpp/tests/reductions/scan_tests.cpp index 2b7faa32316..5abcb5a6402 100644 --- a/cpp/tests/reductions/scan_tests.cpp +++ b/cpp/tests/reductions/scan_tests.cpp @@ -470,6 +470,32 @@ TYPED_TEST(ScanTest, EmptyColumnskip_nulls) CUDF_TEST_EXPECT_COLUMNS_EQUAL(expected_col_out2, col_out->view()); } +TYPED_TEST(ScanTest, LeadingNulls) +{ + bool do_print = true; + auto const v = cudf::test::make_type_param_vector({100, 200, 300}); + auto const b = std::vector{0, 1, 1}; + cudf::test::fixed_width_column_wrapper const col_in(v.begin(), v.end(), b.begin()); + std::unique_ptr col_out; + + // expected outputs + std::vector out_v(v.size()); + std::vector out_b(v.size(), 0); + + // skipna=false + CUDF_EXPECT_NO_THROW( + col_out = + cudf::scan(col_in, cudf::make_sum_aggregation(), scan_type::INCLUSIVE, null_policy::INCLUDE)); + cudf::test::fixed_width_column_wrapper expected_col_out( + out_v.begin(), out_v.end(), out_b.begin()); + if (do_print) { + print_view(expected_col_out, "expect = "); + print_view(col_out->view(), "result = "); + } + CUDF_TEST_EXPECT_COLUMN_PROPERTIES_EQUAL(expected_col_out, col_out->view()); + CUDF_TEST_EXPECT_COLUMNS_EQUAL(expected_col_out, col_out->view()); +} + template struct FixedPointTestBothReps : public cudf::test::BaseFixture { }; From c7f296eced696f973ae01cef78aa6d1fe68a164a Mon Sep 17 00:00:00 2001 From: Wonchan Lee Date: Tue, 23 Feb 2021 13:57:30 -0800 Subject: [PATCH 2/9] Change the mask scan to make the failing test pass --- cpp/src/reductions/scan.cu | 5 +++-- 1 file changed, 3 insertions(+), 2 deletions(-) diff --git a/cpp/src/reductions/scan.cu b/cpp/src/reductions/scan.cu index 011b34031fe..6c42c8aa3f7 100644 --- a/cpp/src/reductions/scan.cu +++ b/cpp/src/reductions/scan.cu @@ -116,8 +116,9 @@ struct scan_dispatcher { thrust::find_if_not( rmm::exec_policy(stream), v, v + input_view.size(), thrust::identity{}) - v; - cudf::set_null_mask( - static_cast(mask.data()), 0, first_null_position, true); + if (first_null_position > 0) + cudf::set_null_mask( + static_cast(mask.data()), 0, first_null_position, true); cudf::set_null_mask( static_cast(mask.data()), first_null_position, input_view.size(), false); return mask; From e19c6d90df747ccd95414b1cfae2e1e864fcc2b4 Mon Sep 17 00:00:00 2001 From: Wonchan Lee Date: Tue, 23 Feb 2021 14:50:48 -0800 Subject: [PATCH 3/9] Update cpp/tests/reductions/scan_tests.cpp Co-authored-by: Vukasin Milovanovic --- cpp/tests/reductions/scan_tests.cpp | 1 - 1 file changed, 1 deletion(-) diff --git a/cpp/tests/reductions/scan_tests.cpp b/cpp/tests/reductions/scan_tests.cpp index 5abcb5a6402..7ee9b81c6d7 100644 --- a/cpp/tests/reductions/scan_tests.cpp +++ b/cpp/tests/reductions/scan_tests.cpp @@ -472,7 +472,6 @@ TYPED_TEST(ScanTest, EmptyColumnskip_nulls) TYPED_TEST(ScanTest, LeadingNulls) { - bool do_print = true; auto const v = cudf::test::make_type_param_vector({100, 200, 300}); auto const b = std::vector{0, 1, 1}; cudf::test::fixed_width_column_wrapper const col_in(v.begin(), v.end(), b.begin()); From 52d2672d76c08941317f95afc6ac0e52b61c7da7 Mon Sep 17 00:00:00 2001 From: Wonchan Lee Date: Tue, 23 Feb 2021 14:51:08 -0800 Subject: [PATCH 4/9] Update cpp/tests/reductions/scan_tests.cpp Co-authored-by: Vukasin Milovanovic --- cpp/tests/reductions/scan_tests.cpp | 4 ---- 1 file changed, 4 deletions(-) diff --git a/cpp/tests/reductions/scan_tests.cpp b/cpp/tests/reductions/scan_tests.cpp index 7ee9b81c6d7..bb8fa8ac079 100644 --- a/cpp/tests/reductions/scan_tests.cpp +++ b/cpp/tests/reductions/scan_tests.cpp @@ -487,10 +487,6 @@ TYPED_TEST(ScanTest, LeadingNulls) cudf::scan(col_in, cudf::make_sum_aggregation(), scan_type::INCLUSIVE, null_policy::INCLUDE)); cudf::test::fixed_width_column_wrapper expected_col_out( out_v.begin(), out_v.end(), out_b.begin()); - if (do_print) { - print_view(expected_col_out, "expect = "); - print_view(col_out->view(), "result = "); - } CUDF_TEST_EXPECT_COLUMN_PROPERTIES_EQUAL(expected_col_out, col_out->view()); CUDF_TEST_EXPECT_COLUMNS_EQUAL(expected_col_out, col_out->view()); } From c334cdd1c221e21a447d3de43e6bb4f719933699 Mon Sep 17 00:00:00 2001 From: Wonchan Lee Date: Tue, 23 Feb 2021 16:28:36 -0800 Subject: [PATCH 5/9] Revert "Change the mask scan to make the failing test pass" This reverts commit c7f296eced696f973ae01cef78aa6d1fe68a164a. --- cpp/src/reductions/scan.cu | 5 ++--- 1 file changed, 2 insertions(+), 3 deletions(-) diff --git a/cpp/src/reductions/scan.cu b/cpp/src/reductions/scan.cu index 6c42c8aa3f7..011b34031fe 100644 --- a/cpp/src/reductions/scan.cu +++ b/cpp/src/reductions/scan.cu @@ -116,9 +116,8 @@ struct scan_dispatcher { thrust::find_if_not( rmm::exec_policy(stream), v, v + input_view.size(), thrust::identity{}) - v; - if (first_null_position > 0) - cudf::set_null_mask( - static_cast(mask.data()), 0, first_null_position, true); + cudf::set_null_mask( + static_cast(mask.data()), 0, first_null_position, true); cudf::set_null_mask( static_cast(mask.data()), first_null_position, input_view.size(), false); return mask; From 55a1cc9d2c39b9d541afc6914ad9eeceb72331e8 Mon Sep 17 00:00:00 2001 From: Wonchan Lee Date: Tue, 23 Feb 2021 16:50:04 -0800 Subject: [PATCH 6/9] Change set_null_mask to handle empty updates correctly --- cpp/src/bitmask/null_mask.cu | 3 ++- 1 file changed, 2 insertions(+), 1 deletion(-) diff --git a/cpp/src/bitmask/null_mask.cu b/cpp/src/bitmask/null_mask.cu index 8cdcefe9796..78188b26473 100644 --- a/cpp/src/bitmask/null_mask.cu +++ b/cpp/src/bitmask/null_mask.cu @@ -135,7 +135,8 @@ void set_null_mask(bitmask_type *bitmask, { CUDF_FUNC_RANGE(); CUDF_EXPECTS(begin_bit >= 0, "Invalid range."); - CUDF_EXPECTS(begin_bit < end_bit, "Invalid bit range."); + CUDF_EXPECTS(begin_bit <= end_bit, "Invalid bit range."); + if (begin_bit == end_bit) return; if (bitmask != nullptr) { auto number_of_mask_words = num_bitmask_words(end_bit) - begin_bit / detail::size_in_bits(); From 3a3eb12761e5e21721bd2a40a9b5cc39627cfa90 Mon Sep 17 00:00:00 2001 From: Wonchan Lee Date: Tue, 23 Feb 2021 23:10:58 -0800 Subject: [PATCH 7/9] Fix to make the style check happy --- cpp/tests/reductions/scan_tests.cpp | 4 ++-- 1 file changed, 2 insertions(+), 2 deletions(-) diff --git a/cpp/tests/reductions/scan_tests.cpp b/cpp/tests/reductions/scan_tests.cpp index bb8fa8ac079..549e5e0d215 100644 --- a/cpp/tests/reductions/scan_tests.cpp +++ b/cpp/tests/reductions/scan_tests.cpp @@ -472,8 +472,8 @@ TYPED_TEST(ScanTest, EmptyColumnskip_nulls) TYPED_TEST(ScanTest, LeadingNulls) { - auto const v = cudf::test::make_type_param_vector({100, 200, 300}); - auto const b = std::vector{0, 1, 1}; + auto const v = cudf::test::make_type_param_vector({100, 200, 300}); + auto const b = std::vector{0, 1, 1}; cudf::test::fixed_width_column_wrapper const col_in(v.begin(), v.end(), b.begin()); std::unique_ptr col_out; From 4d749a384b8c67e77263079fb10ceda89bfd8eba Mon Sep 17 00:00:00 2001 From: Wonchan Lee Date: Wed, 24 Feb 2021 11:18:43 -0800 Subject: [PATCH 8/9] Make bitmask tests consistent with new contracts --- cpp/tests/bitmask/set_nullmask_tests.cu | 7 +++---- 1 file changed, 3 insertions(+), 4 deletions(-) diff --git a/cpp/tests/bitmask/set_nullmask_tests.cu b/cpp/tests/bitmask/set_nullmask_tests.cu index d51336a315f..4c28d2c4c16 100644 --- a/cpp/tests/bitmask/set_nullmask_tests.cu +++ b/cpp/tests/bitmask/set_nullmask_tests.cu @@ -105,8 +105,7 @@ TEST_F(SetBitmaskTest, error_range) std::vector begin_end_fail{ {-1, size}, // begin>=0 {-2, -1}, // begin>=0 - {8, 8}, // begin begin_end_pass{ {0, size}, // begin>=0 {0, 1}, // begin>=0 - {8, 9}, // begin Date: Thu, 25 Feb 2021 15:23:35 -0800 Subject: [PATCH 9/9] Test for the new behavior --- cpp/tests/bitmask/set_nullmask_tests.cu | 1 + 1 file changed, 1 insertion(+) diff --git a/cpp/tests/bitmask/set_nullmask_tests.cu b/cpp/tests/bitmask/set_nullmask_tests.cu index 4c28d2c4c16..ae4896827fd 100644 --- a/cpp/tests/bitmask/set_nullmask_tests.cu +++ b/cpp/tests/bitmask/set_nullmask_tests.cu @@ -115,6 +115,7 @@ TEST_F(SetBitmaskTest, error_range) std::vector begin_end_pass{ {0, size}, // begin>=0 {0, 1}, // begin>=0 + {8, 8}, // begin==end {8, 9}, // begin<=end {size - 1, size}, // begin<=end };