[FEA] Avoid host-side processing in CSV reader
- Dominant language
- C++
- Stars
- 9.8k
- Forks
- 1.1k
- Avg merge
- 3d 6m
- Merged PRs (30d)
- 278
Description
As of 23.10, the CSV reader cannot take advantage of [kvikIO](https://github.com/rapidsai/kvikio) and [GPUDirect Storage](https://developer.nvidia.com/gpudirect-storage) due to several processing steps that require the input data to be present in the system memory. These host-side processing steps include:
1. Decompression;
2. Skipping the partial data row at the start of the byte range (if reading a byte range);
3. Skipping the [byte order mark](https://en.wikipedia.org/wiki/Byte_order_mark) (BOM) chars;
4. Parsing the column names in the header;
Decompression (1) is most likely faster on the CPU (not verified) since all data is compressed in a single block. For CSV files with decompression we may choose to decompress on the host side (also see #5142 and #12255). (2) could readily be processed on the device with `thrust::find` and (3) could be accomplished by a single thread kernel. For (4), we can copy the header data from the device and keep the column name parsing code (also see #12582), which avoids keeping input data on the host at the cost of a small D2H copy.
High level proposal:
When reading compressed CSV files, read to host and decompress, then wrap the host buffer into a datasource and pass to `load_data_and_gather_row_offsets`. When reading uncompressed input, just forward the source to `load_data_and_gather_row_offsets`. There, use `device_read` to load chunks to device memory. Items (2) and (3) can be done here (BOM skipping is currently outside of `load_data_and_gather_row_offsets`).
This approach brings several benefits:
* maintains "byte range" support and avoids loading data outside of the requested byte range
* enables direct device reads for uncompressed inputs
* allows the CSV reader to use kvikIO and increases consistency between IO formats
Contributor guide
Assessment
This issue has not been assessed yet.