Remove specializations and split expensive headers #1415

ahendriksen · 2023-04-14T13:45:29Z

This PR:

Reduces single arch (SM70) build time of libraft.so from 19:30 to 13:00 minutes.
Reduces single arch (SM70) binary size of libraft.so from 232 MB to 163 MB and all-arch from 617MB to 428MB.
Reduces single arch (SM70) rebuild time of libraft.so from 18:54 to 2:28 minutes and number of objects rebuilt from 124 to 32 objects after modification of a single file (touch cpp/include/raft/distance/detail/pairwise_matrix/kernel_sm60.cuh)

This is achieved by following the strategy outlined in issue #1416.

In addition, the following changes have been made:

Isolate raft/logger.hpp so that spdlog is not included during the compilation of most translation units.
Isolate raft::get_pool_memory_resource in anticipation of Move RMM_LOGGING_ASSERT into separate header rmm#1241 to achieve an additional 10s reduction in compilation time per translation unit.
Reduce the veclen instances of the ivf flat scan_interleaved_kernel so that it only takes the maximum value for a datatype or 1. This halves the number of instantiations and should not reduce performance for sophisticated users (who already use powers of two for dataset dimensions). [Discussed with @tfeher]
Added ALL_BENCH and ALL_TESTS CMake targets.

Breaking change:

Removal of raft::distance::MinAndDistanceReduceOp and raft::distance::KVPMinReduce from fused_l2_nn.cuh. They were not used externally (by cudf or cuml for instance), but were used in some tests and benchmarks. References in bench and tests have been replaced with raft::distance::detail::* variants.

tfeher · 2023-04-17T12:35:44Z

Thanks @ahendriksen for this PR! I had only a high level view so far. The savings in compile time and binary size are really great, and having an error message of an accidental instantiation is significant improvement!

ahendriksen · 2023-04-17T13:09:02Z

Thanks @tfeher and apologies for the large PR.

I have tagged this PR as breaking. Officially, this is right, because of the removal of two symbols from fused_l2_nn.cuh. These symbols are not used by any downstream projects though (as confirmed with GitHub code search). Unofficially, this PR breaks downstream because the set of includes for many headers has shrunk. I am keeping track of downstream effects here:

So far, this has surfaced some missing includes. I intend to fix these issues upstream as soon as possible.

ajschmidt8 · 2023-04-17T15:51:41Z

Removing ops-codeowners from the required reviews since it doesn't seem there are any file changes that we're responsible for. Feel free to add us back if necessary.

tfeher

Hi Allard, started to look into the details, here is my first batch of comments

cpp/include/raft/neighbors/ivf_flat-inl.cuh

tfeher · 2023-04-17T20:49:42Z

cpp/include/raft/neighbors/ivf_flat-inl.cuh

+ *
+ * @return the constructed ivf-flat index
+ */
+template <typename T, typename IdxT>


I see that you have changed the the template parameters to be uniformly T and IdxT (earlier we had both value_t, idx_t and T, IdxT pairs). Is that the preferred naming convention @cjnolet?

I mainly did this for consistency. I am fine either way. The T, IdxT convention would be consistent with the naming convention used in the distance APIs.

cpp/include/raft/neighbors/ivf_flat-ext.cuh

cpp/include/raft/neighbors/detail/ivf_flat_interleaved_scan.cuh

cpp/include/raft/neighbors/detail/ivf_flat_interleaved_scan-inl.cuh

cpp/src/distance/detail/kernels/gram_matrix_base_double.cu

cpp/src/distance/detail/pairwise_matrix/dispatch.cu

cpp/src/distance/distance.cu

ahendriksen · 2023-04-17T23:17:24Z

The downstream PRs are building alright now. I have filed two PRs to fix missing includes (one for cuml and one for cugraph).

Some of the test runs break due to runtime linking errors. In the case of cugraph, I believe that this does not indicate a bug. In the case of cuML, I am not sure what is happening.

rapidsai/cugraph#3490: The tests for cuda 11.2 and 11.4 are failing, but this seems to be due to them not picking up the libraft.so build from this PR. The tests for cuda 11.8 are passing, which is expected.

rapidsai/cuml#5363: Reverse situation. Tests for 11.2 and 11.4 are succeeding, but the tests for 11.8 are failing. The test for CUDA 11.8 appears to pick up libraft.so from this PR and libcuml from the integration PR, but the test fail due to missing symbol that has been removed in this PR (raft::distance::pairwise_matrix::instantiation_point). I am not sure what is going on. This is the error:

/opt/conda/envs/test/bin/gtests/libcuml/SG_RIDGE_TEST: symbol lookup error: /opt/conda/envs/test/bin/gtests/libcuml/../../../lib/libcuml++.so: undefined symbol: 
_ZN4raft8distance6detail35pairwise_matrix_instantiation_pointINS1_3ops14l1_distance_opIffiEEiffNS_11identity_opEEEvT_NS1_22pairwise_matrix_paramsIT0_T1_T2_T3_EEP11CUstream_st

~~@cjnolet , @robertmaynard : do you have any suggestions what could cause a mismatch of build headers and binaries in the cuML CI?~~

EDIT: I think I found the issue. Due to a typo, libcuml was built twice: once with RAFT from this PR and once with the 23.06 branch.

Some files use `thrust::nullopt_t` but not do not include `thrust/optional.h` . This PR fixes that. The missing include was surfaced by #3490 This PR is necessary to prevent breakage when rapidsai/raft#1415 is merged. Authors: - Allard Hendriksen (https://github.com/ahendriksen) Approvers: - Chuck Hastings (https://github.com/ChuckHastings) URL: #3493

We define this internally (PRIVATE). In addition, we define RAFT_COMPILED both internally and externally (PUBLIC).

Correlation and Cosine distance both return (1 - similarity) in the pairwise distances apis, meaning that is_min_close is returning the wrong sort order for them. Fix. Authors: - Ben Frederickson (https://github.com/benfred) Approvers: - Corey J. Nolet (https://github.com/cjnolet) URL: rapidsai#1419

Add public functions for reading and writing into individual ivf-pq lists (clusters), in the input space (reconstructed data) and in flat PQ codes. Partially solves (IVF-PQ) rapidsai#1205 Authors: - Artem M. Chirkin (https://github.com/achirkin) - Corey J. Nolet (https://github.com/cjnolet) - Tamas Bela Feher (https://github.com/tfeher) Approvers: - Corey J. Nolet (https://github.com/cjnolet) URL: rapidsai#1298

Authors: - Ben Frederickson (https://github.com/benfred) Approvers: - Corey J. Nolet (https://github.com/cjnolet) URL: rapidsai#1422

ahendriksen · 2023-04-18T13:05:28Z

Status update:

Review comments up till now have been addressed
[WIP] Test RAFT PR 1441 cugraph#3490
- Build + tests are in good shape (only failing pre-CUDA-11.8)
- Missing includes are merged: [ENH] Add missing include of thrust/optional.h cugraph#3493
[WIP] Test changes in RAFT PR 1441 cuml#5363:
- Build + tests are in good shape (only failing pre-CUDA-11.8)
- Missing includes have been merged: [ENH] Add missing includes of rmm/mr/device/per_device_resource.hpp cuml#5369

Compiling test/distance/gram.cu would fail otherwise, as it did not explicitly instantiate the raft::distance::distance.

Some files use `rmm::mr::get_current_device_resource()` but not do not include `rmm/mr/device/per_device_resource.hpp` . This PR fixes that. The missing include was surfaced by #5363. This PR is necessary to prevent breakage when rapidsai/raft#1415 is merged. Authors: - Allard Hendriksen (https://github.com/ahendriksen) Approvers: - Corey J. Nolet (https://github.com/cjnolet) URL: #5369

tfeher

Hi Allard, I went through all the changes. On one hand this looks great: the bulk of the changes are just trivial refactoring, and it is great to have everything in one place to see the combined benefit on compilation time, and to test its effects downstream.

On the other hand, the size of the PR makes it tedious even to scroll through the files. That is not an issue in itself, but I fear that might slow down its acceptance. I think the points in the "additional changes" in the PR description would deserve their own PR, to have a focused discussion. Additionally, if we would need to revert any of the changes, then we could do that more targeted. Since I have already went through the PR, I do not insist on separating it, but it could be still beneficial to do so.

cpp/include/raft/core/logger-ext.hpp

cpp/include/raft/distance/distance-ext.cuh

cpp/include/raft/distance/fused_l2_nn-inl.cuh

tfeher · 2023-04-18T22:59:43Z

cpp/include/raft/neighbors/brute_force-ext.cuh

+ epilogue_op distance_epilogue);
+
+instantiate_raft_neighbors_brute_force_knn(
+ int64_t, float, uint32_t, raft::row_major, raft::row_major, raft::identity_op);


Similar question as for the distance prims: does it make sense to have instantiations for both uint32_t and int64_t as matrix_idx?

I don't think it makes sense. The instantiations reflect the current state of things. We need these to make sure libraft.so compiles and the tests compile. I could create an issue to look into the instantiations of brute_force_knn?

That sounds good to create an issue to track these things. Indeed, we want to keep the current state.

Filed issue #1449

cpp/src/neighbors/brute_force_fused_l2_knn_float_int64_t.cu

tfeher · 2023-04-19T07:56:29Z

cpp/test/neighbors/ann_ivf_pq/test_float_uint32_t.cu

+// XXX: the uint32_t instance is not compiled in libraft.so. So we allow
+// instantiating the template here.
+//
+// TODO: consider removing this test or consider adding an instantiation to the


While I understand the benefit of testing for additional idx type, I do not like the fact that this test does not use precompiled insatnces. This leads to either slower development workflow or forces the user to comment out this test for quicker iteration (e.g. while adding / modifying test).

Filed issue #1448

cpp/include/raft/distance/detail/kernels/kernel_matrices.cuh

tfeher · 2023-04-19T12:35:05Z

For new developers, it might be not obvious when do we use the header.cuh / header-inl.cuh / header-ext.cuh triplet instead of just header.cuh. It would be good to add a section to the developer guide (around https://github.com/rapidsai/raft/blob/branch-23.06/docs/source/developer_guide.md#common-design-considerations), something like the header organization section in #1416.

Only keep in logger-ext.hpp: I don't know how to document the logger class otherwise.

ahendriksen · 2023-04-19T17:42:27Z

For new developers, it might be not obvious when do we use the header.cuh / header-inl.cuh / header-ext.cuh triplet instead of just header.cuh. It would be good to add a section to the developer guide

Agreed. I have added a section there.

ahendriksen · 2023-04-20T19:00:01Z

I have split up this PR in smaller pieces:

This should be easier to review. I am leaning towards closing this PR.

github-actions bot added CMake cpp labels Apr 14, 2023

ahendriksen mentioned this pull request Apr 14, 2023

[DISCUSSION] A robust solution to precompiling function templates #1416

Closed

github-actions bot added the python label Apr 14, 2023

ahendriksen added improvement Improvement / enhancement to an existing function Build Time Improvement labels Apr 14, 2023

This was referenced Apr 17, 2023

[WIP] Test changes in RAFT PR 1441 rapidsai/cuml#5363

Closed

[WIP] Test RAFT PR 1415 rapidsai/cudf#13147

Closed

[WIP] Test RAFT PR 1441 rapidsai/cugraph#3490

Closed

ahendriksen marked this pull request as ready for review April 17, 2023 12:13

ahendriksen requested review from a team as code owners April 17, 2023 12:13

ahendriksen added the breaking Breaking change label Apr 17, 2023

ahendriksen mentioned this pull request Apr 17, 2023

[ENH] Add missing include of thrust/optional.h rapidsai/cugraph#3493

Merged

ajschmidt8 removed the request for review from a team April 17, 2023 15:49

tfeher requested changes Apr 17, 2023

View reviewed changes

ahendriksen mentioned this pull request Apr 17, 2023

[ENH] Add missing includes of rmm/mr/device/per_device_resource.hpp rapidsai/cuml#5369

Merged

ahendriksen added 6 commits April 18, 2023 12:20

Move headers to -inl path

65ec2f1

Add back empty headers

51e8673

Disable warnings for clang compilation

9e03fca

Comment out omp to enable clang compilation

3bace7b

cmake: Define RAFT_EXPLICIT_INCLUDE

f0e74a2

We define this internally (PRIVATE). In addition, we define RAFT_COMPILED both internally and externally (PUBLIC).

Split raft/core/logger

a8f62fb

benfred and others added 6 commits April 18, 2023 14:25

Add python bindings for matrix::select_k (rapidsai#1422)

5e84daa

Authors: - Ben Frederickson (https://github.com/benfred) Approvers: - Corey J. Nolet (https://github.com/cjnolet) URL: rapidsai#1422

Fixup yaml and toml files

9906aba

Fixup ivpq and ivf_flat files

1299b23

Merge remote-tracking branch 'rapids/branch-23.06' into enh-inl-headers

868ada3

ahendriksen force-pushed the enh-inl-headers branch from bfdf9fb to 868ada3 Compare April 18, 2023 12:49

Fix style

53f23fb

tfeher mentioned this pull request Apr 18, 2023

CAGRA template instantiations #1428

Closed

distance: Add back rbf_fin_op instances

c684a15

Compiling test/distance/gram.cu would fail otherwise, as it did not explicitly instantiate the raft::distance::distance.

tfeher requested changes Apr 19, 2023

View reviewed changes

ahendriksen added 3 commits April 19, 2023 14:15

Remove unused code

6a6b1e5

Remove spurious macros from brute_force instances

486c2e9

Add documentation for rbf_fin_op

fe6d335

ahendriksen added 7 commits April 19, 2023 15:13

Add documentation for split header structure

be030b8

Rename RAFT_EXPLICIT_INSTANTIATE => RAFT_EXPLICIT_INSTANTIATE_ONLY

eb0d3b2

Remove docstrings from -ext headers

0b9b796

Only keep in logger-ext.hpp: I don't know how to document the logger class otherwise.

Fix extraneous and missing includes in headers

e8fc2da

Add tests to ensure headers compile in isolation

488a273

Add macro tables to developer guide

f37c0e7

Merge remote-tracking branch 'rapids/branch-23.06' into enh-inl-headers

1b899f9

ahendriksen mentioned this pull request Apr 19, 2023

[PoC] select_k: Replace specialization by split header #1434

Closed

ahendriksen mentioned this pull request Apr 20, 2023

[ENH] [1/5] Header structure: replace specializations #1437

Closed

ahendriksen mentioned this pull request Apr 21, 2023

[FEA] Consider removing uint32_t instance of raft::neighbors::brute_force::knn #1449

Open

cjnolet closed this Apr 28, 2023

Provide feedback

Saved searches

Use saved searches to filter your results more quickly

Remove specializations and split expensive headers #1415

Remove specializations and split expensive headers #1415

ahendriksen commented Apr 14, 2023 •

edited

Loading

tfeher commented Apr 17, 2023

ahendriksen commented Apr 17, 2023

ajschmidt8 commented Apr 17, 2023

tfeher left a comment

tfeher Apr 17, 2023

ahendriksen Apr 18, 2023

ahendriksen commented Apr 17, 2023 •

edited

Loading

ahendriksen commented Apr 18, 2023 •

edited

Loading

tfeher left a comment

tfeher Apr 18, 2023

ahendriksen Apr 19, 2023

tfeher Apr 19, 2023

ahendriksen Apr 21, 2023

tfeher Apr 19, 2023

ahendriksen Apr 21, 2023

tfeher commented Apr 19, 2023

ahendriksen commented Apr 19, 2023

ahendriksen commented Apr 20, 2023

Remove specializations and split expensive headers #1415

Remove specializations and split expensive headers #1415

Conversation

ahendriksen commented Apr 14, 2023 • edited Loading

tfeher commented Apr 17, 2023

ahendriksen commented Apr 17, 2023

ajschmidt8 commented Apr 17, 2023

tfeher left a comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

ahendriksen commented Apr 17, 2023 • edited Loading

ahendriksen commented Apr 18, 2023 • edited Loading

tfeher left a comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

tfeher commented Apr 19, 2023

ahendriksen commented Apr 19, 2023

ahendriksen commented Apr 20, 2023

ahendriksen commented Apr 14, 2023 •

edited

Loading

ahendriksen commented Apr 17, 2023 •

edited

Loading

ahendriksen commented Apr 18, 2023 •

edited

Loading