From 743bcf0cd18ffac154634ca757a596d74da45434 Mon Sep 17 00:00:00 2001 From: seidl Date: Mon, 13 Mar 2023 15:55:17 -0700 Subject: [PATCH 1/6] fix calculation of null values for parquet stats --- cpp/src/io/parquet/page_enc.cu | 1 + cpp/src/io/statistics/column_statistics.cuh | 10 +- cpp/src/io/statistics/statistics.cuh | 3 +- cpp/tests/io/parquet_test.cpp | 138 ++++++++++++++++++++ 4 files changed, 149 insertions(+), 3 deletions(-) diff --git a/cpp/src/io/parquet/page_enc.cu b/cpp/src/io/parquet/page_enc.cu index 5a12acec2a3..e48696fcb9b 100644 --- a/cpp/src/io/parquet/page_enc.cu +++ b/cpp/src/io/parquet/page_enc.cu @@ -282,6 +282,7 @@ __global__ void __launch_bounds__(128) g.col = ck_g->col_desc; g.start_row = fragments[frag_id].start_value_idx; g.num_rows = fragments[frag_id].num_leaf_values; + g.non_leaf_nulls = fragments[frag_id].num_values - g.num_rows; groups[frag_id] = g; } } diff --git a/cpp/src/io/statistics/column_statistics.cuh b/cpp/src/io/statistics/column_statistics.cuh index 125235ebf2f..0b09cb63d19 100644 --- a/cpp/src/io/statistics/column_statistics.cuh +++ b/cpp/src/io/statistics/column_statistics.cuh @@ -1,5 +1,5 @@ /* - * Copyright (c) 2021-2022, NVIDIA CORPORATION. + * Copyright (c) 2021-2023, NVIDIA CORPORATION. * * Licensed under the Apache License, Version 2.0 (the "License"); * you may not use this file except in compliance with the License. @@ -129,7 +129,13 @@ struct calculate_group_statistics_functor { chunk = block_reduce(chunk, storage); - if (t == 0) { s.ck = get_untyped_chunk(chunk); } + if (t == 0) { + // parquet wants total null count in stats, not just count of null leaf values + if constexpr (IO == detail::io_file_format::PARQUET) { + chunk.null_count += s.group.non_leaf_nulls; + } + s.ck = get_untyped_chunk(chunk); + } } template ; + + // 4 nulls + // [NULL, 2, NULL] + // [] + // [4, 5] + // NULL + lcw col0{{{{1, 2, 3}, valids}, {}, {4, 5}, {}}, valids2}; + + // 4 nulls + // [[1, 2, 3], [], [4, 5], [], [0, 6, 0]] + // [[7, 8]] + // [] + // [[]] + lcw col1{{{1, 2, 3}, {}, {4, 5}, {}, {0, 6, 0}}, {{7, 8}}, lcw{}, lcw{lcw{}}}; + + // 4 nulls + // [[1, 2, 3], [], [4, 5], NULL, [0, 6, 0]] + // [[7, 8]] + // [] + // [[]] + lcw col2{{{{1, 2, 3}, {}, {4, 5}, {}, {0, 6, 0}}, valids2}, {{7, 8}}, lcw{}, lcw{lcw{}}}; + + // 6 nulls + // [[1, 2, 3], [], [4, 5], NULL, [NULL, 6, NULL]] + // [[7, 8]] + // [] + // [[]] + using dlcw = cudf::test::lists_column_wrapper; + dlcw col3{{{{1., 2., 3.}, {}, {4., 5.}, {}, {{0., 6., 0.}, valids}}, valids2}, + {{7., 8.}}, + dlcw{}, + dlcw{dlcw{}}}; + + // 4 nulls + // TODO: uint16_t lists are not read properly in parquet reader + // [[1, 2, 3], [], [4, 5], NULL, [0, 6, 0]] + // [[7, 8]] + // [] + // NULL + using ui16lcw = cudf::test::lists_column_wrapper; + cudf::test::lists_column_wrapper col4{ + {{{{1, 2, 3}, {}, {4, 5}, {}, {0, 6, 0}}, valids2}, {{7, 8}}, ui16lcw{}, ui16lcw{ui16lcw{}}}, + valids2}; + + // 6 nulls + // [[1, 2, 3], [], [4, 5], NULL, [NULL, 6, NULL]] + // [[7, 8]] + // [] + // NULL + lcw col5{ + {{{{1, 2, 3}, {}, {4, 5}, {}, {{0, 6, 0}, valids}}, valids2}, {{7, 8}}, lcw{}, lcw{lcw{}}}, + valids2}; + + // 4 nulls + using strlcw = cudf::test::lists_column_wrapper; + cudf::test::lists_column_wrapper col6{ + {{"Monday", "Monday", "Friday"}, {}, {"Monday", "Friday"}, {}, {"Sunday", "Funday"}}, + {{"bee", "sting"}}, + strlcw{}, + strlcw{strlcw{}}}; + + // 11 nulls + // [[[NULL,2,NULL,4]], [[NULL,6,NULL], [8,9]]] + // [NULL, [[13],[14,15,16]], NULL] + // [NULL, [], NULL, [[]]] + // NULL + lcw col7{{ + {{{{1, 2, 3, 4}, valids}}, {{{5, 6, 7}, valids}, {8, 9}}}, + {{{{10, 11}, {12}}, {{13}, {14, 15, 16}}, {{17, 18}}}, valids}, + {{lcw{lcw{}}, lcw{}, lcw{}, lcw{lcw{}}}, valids}, + lcw{lcw{lcw{}}}, + }, + valids2}; + + table_view expected({col0, col1, col2, col3, col4, col5, col6, col7}); + + cudf::io::table_input_metadata expected_metadata(expected); + expected_metadata.column_metadata[0].set_name("col_list_int_0"); + expected_metadata.column_metadata[1].set_name("col_list_list_int_1"); + expected_metadata.column_metadata[2].set_name("col_list_list_int_nullable_2"); + expected_metadata.column_metadata[3].set_name("col_list_list_nullable_double_nullable_3"); + expected_metadata.column_metadata[0].set_name("col_list_list_uint16_4"); + expected_metadata.column_metadata[4].set_name("col_list_nullable_list_nullable_int_nullable_5"); + expected_metadata.column_metadata[5].set_name("col_list_list_string_6"); + expected_metadata.column_metadata[6].set_name("col_list_list_list_7"); + + int64_t const null_counts[] = {4, 4, 4, 6, 4, 6, 4, 11}; + + auto const filepath = temp_env->get_temp_filepath("ColumnIndexListWithNulls.parquet"); + auto out_opts = cudf::io::parquet_writer_options::builder(cudf::io::sink_info{filepath}, expected) + .metadata(&expected_metadata) + .stats_level(cudf::io::statistics_freq::STATISTICS_COLUMN) + .compression(cudf::io::compression_type::NONE); + + cudf::io::write_parquet(out_opts); + + auto const source = cudf::io::datasource::create(filepath); + cudf::io::parquet::FileMetaData fmd; + + read_footer(source, &fmd); + + for (size_t r = 0; r < fmd.row_groups.size(); r++) { + auto const& rg = fmd.row_groups[r]; + for (size_t c = 0; c < rg.columns.size(); c++) { + auto const& chunk = rg.columns[c]; + + // loop over offsets, read each page header, make sure it's a data page and that + // the first row index is correct + auto const oi = read_offset_index(source, chunk); + + int64_t num_vals = 0; + for (size_t o = 0; o < oi.page_locations.size(); o++) { + auto const& page_loc = oi.page_locations[o]; + auto const ph = read_page_header(source, page_loc); + EXPECT_EQ(ph.type, cudf::io::parquet::PageType::DATA_PAGE); + // last column has 2 values per row + EXPECT_EQ(page_loc.first_row_index * (c == rg.columns.size() - 1 ? 2 : 1), num_vals); + num_vals += ph.data_page_header.num_values; + } + + // check null counts in column chunk stats and page indexes + auto const ci = read_column_index(source, chunk); + auto const stats = parse_statistics(chunk); + EXPECT_EQ(stats.null_count, null_counts[c]); + + // should only be one page + EXPECT_FALSE(ci.null_pages[0]); + EXPECT_EQ(ci.null_counts[0], null_counts[c]); + } + } +} + TEST_F(ParquetWriterTest, CheckColumnIndexTruncation) { const char* coldata[] = { From 41604abad1d30afdad6110bf472ced04b49ac485 Mon Sep 17 00:00:00 2001 From: seidl Date: Mon, 13 Mar 2023 16:09:34 -0700 Subject: [PATCH 2/6] check nulls in more tests --- cpp/tests/io/parquet_test.cpp | 6 ++++++ 1 file changed, 6 insertions(+) diff --git a/cpp/tests/io/parquet_test.cpp b/cpp/tests/io/parquet_test.cpp index 5dcaf1e5990..b2f3f55ae5a 100644 --- a/cpp/tests/io/parquet_test.cpp +++ b/cpp/tests/io/parquet_test.cpp @@ -4279,6 +4279,9 @@ TEST_F(ParquetWriterTest, CheckColumnOffsetIndexNulls) auto const ci = read_column_index(source, chunk); auto const stats = parse_statistics(chunk); + // should be half nulls, except no nulls in column 0 + EXPECT_EQ(stats.null_count, c > 0 ? num_rows / 2 : 0); + // schema indexing starts at 1 auto const ptype = fmd.schema[c + 1].type; auto const ctype = fmd.schema[c + 1].converted_type; @@ -4364,6 +4367,9 @@ TEST_F(ParquetWriterTest, CheckColumnOffsetIndexNullColumn) auto const ci = read_column_index(source, chunk); auto const stats = parse_statistics(chunk); + // there should be no nulls except column 1 which is all nulls + EXPECT_EQ(stats.null_count, c == 1 ? num_rows : 0); + // schema indexing starts at 1 auto const ptype = fmd.schema[c + 1].type; auto const ctype = fmd.schema[c + 1].converted_type; From 313be05afdaa0dd094c404adec724b26f16bf7b7 Mon Sep 17 00:00:00 2001 From: seidl Date: Mon, 13 Mar 2023 16:13:54 -0700 Subject: [PATCH 3/6] remove unneeded metadata --- cpp/tests/io/parquet_test.cpp | 11 ----------- 1 file changed, 11 deletions(-) diff --git a/cpp/tests/io/parquet_test.cpp b/cpp/tests/io/parquet_test.cpp index b2f3f55ae5a..824f9dbf67d 100644 --- a/cpp/tests/io/parquet_test.cpp +++ b/cpp/tests/io/parquet_test.cpp @@ -4553,21 +4553,10 @@ TEST_F(ParquetWriterTest, CheckColumnIndexListWithNulls) table_view expected({col0, col1, col2, col3, col4, col5, col6, col7}); - cudf::io::table_input_metadata expected_metadata(expected); - expected_metadata.column_metadata[0].set_name("col_list_int_0"); - expected_metadata.column_metadata[1].set_name("col_list_list_int_1"); - expected_metadata.column_metadata[2].set_name("col_list_list_int_nullable_2"); - expected_metadata.column_metadata[3].set_name("col_list_list_nullable_double_nullable_3"); - expected_metadata.column_metadata[0].set_name("col_list_list_uint16_4"); - expected_metadata.column_metadata[4].set_name("col_list_nullable_list_nullable_int_nullable_5"); - expected_metadata.column_metadata[5].set_name("col_list_list_string_6"); - expected_metadata.column_metadata[6].set_name("col_list_list_list_7"); - int64_t const null_counts[] = {4, 4, 4, 6, 4, 6, 4, 11}; auto const filepath = temp_env->get_temp_filepath("ColumnIndexListWithNulls.parquet"); auto out_opts = cudf::io::parquet_writer_options::builder(cudf::io::sink_info{filepath}, expected) - .metadata(&expected_metadata) .stats_level(cudf::io::statistics_freq::STATISTICS_COLUMN) .compression(cudf::io::compression_type::NONE); From 2a9b9279780fa6a6ec2284b504779e957e439f32 Mon Sep 17 00:00:00 2001 From: seidl Date: Mon, 13 Mar 2023 16:15:18 -0700 Subject: [PATCH 4/6] remove todo --- cpp/tests/io/parquet_test.cpp | 1 - 1 file changed, 1 deletion(-) diff --git a/cpp/tests/io/parquet_test.cpp b/cpp/tests/io/parquet_test.cpp index 824f9dbf67d..ebc3c5feb1b 100644 --- a/cpp/tests/io/parquet_test.cpp +++ b/cpp/tests/io/parquet_test.cpp @@ -4511,7 +4511,6 @@ TEST_F(ParquetWriterTest, CheckColumnIndexListWithNulls) dlcw{dlcw{}}}; // 4 nulls - // TODO: uint16_t lists are not read properly in parquet reader // [[1, 2, 3], [], [4, 5], NULL, [0, 6, 0]] // [[7, 8]] // [] From 42b57e1db286154ea24f8d95e69da8ee30a5d8aa Mon Sep 17 00:00:00 2001 From: seidl Date: Tue, 14 Mar 2023 14:46:25 -0700 Subject: [PATCH 5/6] descriptive names for validity iterators --- cpp/tests/io/parquet_test.cpp | 40 ++++++++++++++++++++--------------- 1 file changed, 23 insertions(+), 17 deletions(-) diff --git a/cpp/tests/io/parquet_test.cpp b/cpp/tests/io/parquet_test.cpp index ebc3c5feb1b..b649cd507af 100644 --- a/cpp/tests/io/parquet_test.cpp +++ b/cpp/tests/io/parquet_test.cpp @@ -4473,8 +4473,9 @@ TEST_F(ParquetWriterTest, CheckColumnOffsetIndexStruct) TEST_F(ParquetWriterTest, CheckColumnIndexListWithNulls) { - auto valids = cudf::detail::make_counting_transform_iterator(0, [](auto i) { return i % 2; }); - auto valids2 = cudf::detail::make_counting_transform_iterator(0, [](auto i) { return i != 3; }); + auto null_at_even_idx = + cudf::detail::make_counting_transform_iterator(0, [](auto i) { return i % 2; }); + auto null_at_idx_3 = cudf::test::iterators::null_at(3); using lcw = cudf::test::lists_column_wrapper; @@ -4483,7 +4484,7 @@ TEST_F(ParquetWriterTest, CheckColumnIndexListWithNulls) // [] // [4, 5] // NULL - lcw col0{{{{1, 2, 3}, valids}, {}, {4, 5}, {}}, valids2}; + lcw col0{{{{1, 2, 3}, null_at_even_idx}, {}, {4, 5}, {}}, null_at_idx_3}; // 4 nulls // [[1, 2, 3], [], [4, 5], [], [0, 6, 0]] @@ -4497,7 +4498,7 @@ TEST_F(ParquetWriterTest, CheckColumnIndexListWithNulls) // [[7, 8]] // [] // [[]] - lcw col2{{{{1, 2, 3}, {}, {4, 5}, {}, {0, 6, 0}}, valids2}, {{7, 8}}, lcw{}, lcw{lcw{}}}; + lcw col2{{{{1, 2, 3}, {}, {4, 5}, {}, {0, 6, 0}}, null_at_idx_3}, {{7, 8}}, lcw{}, lcw{lcw{}}}; // 6 nulls // [[1, 2, 3], [], [4, 5], NULL, [NULL, 6, NULL]] @@ -4505,7 +4506,7 @@ TEST_F(ParquetWriterTest, CheckColumnIndexListWithNulls) // [] // [[]] using dlcw = cudf::test::lists_column_wrapper; - dlcw col3{{{{1., 2., 3.}, {}, {4., 5.}, {}, {{0., 6., 0.}, valids}}, valids2}, + dlcw col3{{{{1., 2., 3.}, {}, {4., 5.}, {}, {{0., 6., 0.}, null_at_even_idx}}, null_at_idx_3}, {{7., 8.}}, dlcw{}, dlcw{dlcw{}}}; @@ -4517,17 +4518,22 @@ TEST_F(ParquetWriterTest, CheckColumnIndexListWithNulls) // NULL using ui16lcw = cudf::test::lists_column_wrapper; cudf::test::lists_column_wrapper col4{ - {{{{1, 2, 3}, {}, {4, 5}, {}, {0, 6, 0}}, valids2}, {{7, 8}}, ui16lcw{}, ui16lcw{ui16lcw{}}}, - valids2}; + {{{{1, 2, 3}, {}, {4, 5}, {}, {0, 6, 0}}, null_at_idx_3}, + {{7, 8}}, + ui16lcw{}, + ui16lcw{ui16lcw{}}}, + null_at_idx_3}; // 6 nulls // [[1, 2, 3], [], [4, 5], NULL, [NULL, 6, NULL]] // [[7, 8]] // [] // NULL - lcw col5{ - {{{{1, 2, 3}, {}, {4, 5}, {}, {{0, 6, 0}, valids}}, valids2}, {{7, 8}}, lcw{}, lcw{lcw{}}}, - valids2}; + lcw col5{{{{{1, 2, 3}, {}, {4, 5}, {}, {{0, 6, 0}, null_at_even_idx}}, null_at_idx_3}, + {{7, 8}}, + lcw{}, + lcw{lcw{}}}, + null_at_idx_3}; // 4 nulls using strlcw = cudf::test::lists_column_wrapper; @@ -4543,16 +4549,16 @@ TEST_F(ParquetWriterTest, CheckColumnIndexListWithNulls) // [NULL, [], NULL, [[]]] // NULL lcw col7{{ - {{{{1, 2, 3, 4}, valids}}, {{{5, 6, 7}, valids}, {8, 9}}}, - {{{{10, 11}, {12}}, {{13}, {14, 15, 16}}, {{17, 18}}}, valids}, - {{lcw{lcw{}}, lcw{}, lcw{}, lcw{lcw{}}}, valids}, + {{{{1, 2, 3, 4}, null_at_even_idx}}, {{{5, 6, 7}, null_at_even_idx}, {8, 9}}}, + {{{{10, 11}, {12}}, {{13}, {14, 15, 16}}, {{17, 18}}}, null_at_even_idx}, + {{lcw{lcw{}}, lcw{}, lcw{}, lcw{lcw{}}}, null_at_even_idx}, lcw{lcw{lcw{}}}, }, - valids2}; + null_at_idx_3}; table_view expected({col0, col1, col2, col3, col4, col5, col6, col7}); - int64_t const null_counts[] = {4, 4, 4, 6, 4, 6, 4, 11}; + int64_t const expected_null_counts[] = {4, 4, 4, 6, 4, 6, 4, 11}; auto const filepath = temp_env->get_temp_filepath("ColumnIndexListWithNulls.parquet"); auto out_opts = cudf::io::parquet_writer_options::builder(cudf::io::sink_info{filepath}, expected) @@ -4588,11 +4594,11 @@ TEST_F(ParquetWriterTest, CheckColumnIndexListWithNulls) // check null counts in column chunk stats and page indexes auto const ci = read_column_index(source, chunk); auto const stats = parse_statistics(chunk); - EXPECT_EQ(stats.null_count, null_counts[c]); + EXPECT_EQ(stats.null_count, expected_null_counts[c]); // should only be one page EXPECT_FALSE(ci.null_pages[0]); - EXPECT_EQ(ci.null_counts[0], null_counts[c]); + EXPECT_EQ(ci.null_counts[0], expected_null_counts[c]); } } } From c7dd41c87b245358ae4e184f61587355e231b3a9 Mon Sep 17 00:00:00 2001 From: seidl Date: Tue, 14 Mar 2023 15:58:02 -0700 Subject: [PATCH 6/6] implement suggestions from review --- cpp/tests/io/parquet_test.cpp | 33 ++++++++++++++------------------- 1 file changed, 14 insertions(+), 19 deletions(-) diff --git a/cpp/tests/io/parquet_test.cpp b/cpp/tests/io/parquet_test.cpp index b649cd507af..cf94a0877d2 100644 --- a/cpp/tests/io/parquet_test.cpp +++ b/cpp/tests/io/parquet_test.cpp @@ -4280,7 +4280,7 @@ TEST_F(ParquetWriterTest, CheckColumnOffsetIndexNulls) auto const stats = parse_statistics(chunk); // should be half nulls, except no nulls in column 0 - EXPECT_EQ(stats.null_count, c > 0 ? num_rows / 2 : 0); + EXPECT_EQ(stats.null_count, c == 0 ? 0 : num_rows / 2); // schema indexing starts at 1 auto const ptype = fmd.schema[c + 1].type; @@ -4473,10 +4473,8 @@ TEST_F(ParquetWriterTest, CheckColumnOffsetIndexStruct) TEST_F(ParquetWriterTest, CheckColumnIndexListWithNulls) { - auto null_at_even_idx = - cudf::detail::make_counting_transform_iterator(0, [](auto i) { return i % 2; }); - auto null_at_idx_3 = cudf::test::iterators::null_at(3); - + using cudf::test::iterators::null_at; + using cudf::test::iterators::nulls_at; using lcw = cudf::test::lists_column_wrapper; // 4 nulls @@ -4484,7 +4482,7 @@ TEST_F(ParquetWriterTest, CheckColumnIndexListWithNulls) // [] // [4, 5] // NULL - lcw col0{{{{1, 2, 3}, null_at_even_idx}, {}, {4, 5}, {}}, null_at_idx_3}; + lcw col0{{{{1, 2, 3}, nulls_at({0, 2})}, {}, {4, 5}, {}}, null_at(3)}; // 4 nulls // [[1, 2, 3], [], [4, 5], [], [0, 6, 0]] @@ -4498,7 +4496,7 @@ TEST_F(ParquetWriterTest, CheckColumnIndexListWithNulls) // [[7, 8]] // [] // [[]] - lcw col2{{{{1, 2, 3}, {}, {4, 5}, {}, {0, 6, 0}}, null_at_idx_3}, {{7, 8}}, lcw{}, lcw{lcw{}}}; + lcw col2{{{{1, 2, 3}, {}, {4, 5}, {}, {0, 6, 0}}, null_at(3)}, {{7, 8}}, lcw{}, lcw{lcw{}}}; // 6 nulls // [[1, 2, 3], [], [4, 5], NULL, [NULL, 6, NULL]] @@ -4506,7 +4504,7 @@ TEST_F(ParquetWriterTest, CheckColumnIndexListWithNulls) // [] // [[]] using dlcw = cudf::test::lists_column_wrapper; - dlcw col3{{{{1., 2., 3.}, {}, {4., 5.}, {}, {{0., 6., 0.}, null_at_even_idx}}, null_at_idx_3}, + dlcw col3{{{{1., 2., 3.}, {}, {4., 5.}, {}, {{0., 6., 0.}, nulls_at({0, 2})}}, null_at(3)}, {{7., 8.}}, dlcw{}, dlcw{dlcw{}}}; @@ -4518,22 +4516,19 @@ TEST_F(ParquetWriterTest, CheckColumnIndexListWithNulls) // NULL using ui16lcw = cudf::test::lists_column_wrapper; cudf::test::lists_column_wrapper col4{ - {{{{1, 2, 3}, {}, {4, 5}, {}, {0, 6, 0}}, null_at_idx_3}, - {{7, 8}}, - ui16lcw{}, - ui16lcw{ui16lcw{}}}, - null_at_idx_3}; + {{{{1, 2, 3}, {}, {4, 5}, {}, {0, 6, 0}}, null_at(3)}, {{7, 8}}, ui16lcw{}, ui16lcw{ui16lcw{}}}, + null_at(3)}; // 6 nulls // [[1, 2, 3], [], [4, 5], NULL, [NULL, 6, NULL]] // [[7, 8]] // [] // NULL - lcw col5{{{{{1, 2, 3}, {}, {4, 5}, {}, {{0, 6, 0}, null_at_even_idx}}, null_at_idx_3}, + lcw col5{{{{{1, 2, 3}, {}, {4, 5}, {}, {{0, 6, 0}, nulls_at({0, 2})}}, null_at(3)}, {{7, 8}}, lcw{}, lcw{lcw{}}}, - null_at_idx_3}; + null_at(3)}; // 4 nulls using strlcw = cudf::test::lists_column_wrapper; @@ -4549,12 +4544,12 @@ TEST_F(ParquetWriterTest, CheckColumnIndexListWithNulls) // [NULL, [], NULL, [[]]] // NULL lcw col7{{ - {{{{1, 2, 3, 4}, null_at_even_idx}}, {{{5, 6, 7}, null_at_even_idx}, {8, 9}}}, - {{{{10, 11}, {12}}, {{13}, {14, 15, 16}}, {{17, 18}}}, null_at_even_idx}, - {{lcw{lcw{}}, lcw{}, lcw{}, lcw{lcw{}}}, null_at_even_idx}, + {{{{1, 2, 3, 4}, nulls_at({0, 2})}}, {{{5, 6, 7}, nulls_at({0, 2})}, {8, 9}}}, + {{{{10, 11}, {12}}, {{13}, {14, 15, 16}}, {{17, 18}}}, nulls_at({0, 2})}, + {{lcw{lcw{}}, lcw{}, lcw{}, lcw{lcw{}}}, nulls_at({0, 2})}, lcw{lcw{lcw{}}}, }, - null_at_idx_3}; + null_at(3)}; table_view expected({col0, col1, col2, col3, col4, col5, col6, col7});