diff --git a/INSTALL.md b/INSTALL.md index 9d928a4ea..80c6ec9a2 100644 --- a/INSTALL.md +++ b/INSTALL.md @@ -31,6 +31,17 @@ $ conda install -c pytorch/label/nightly faiss-cpu $ conda install -c pytorch/label/nightly faiss-gpu ``` +A combination of versions that works with Pytorch (as of 2022-11-23): +``` +conda create -n faiss_1.7.3 python=3.8 +conda activate faiss_1.7.3 +conda install pytorch==1.11.0 cudatoolkit=11.3 -c pytorch +conda install numpy +conda install -c pytorch faiss-gpu=1.7.3 cudatoolkit=11.3 +conda install -c conda-forge notebook +conda install -y matplotlib +``` + ## Installing from conda-forge Faiss is also being packaged by [conda-forge](https://conda-forge.org/), the diff --git a/README.md b/README.md index f76669fe2..5c6f6e453 100644 --- a/README.md +++ b/README.md @@ -1,6 +1,6 @@ # Faiss -Faiss is a library for efficient similarity search and clustering of dense vectors. It contains algorithms that search in sets of vectors of any size, up to ones that possibly do not fit in RAM. It also contains supporting code for evaluation and parameter tuning. Faiss is written in C++ with complete wrappers for Python/numpy. Some of the most useful algorithms are implemented on the GPU. It is developed primarily at [Facebook AI Research](https://ai.facebook.com/). +Faiss is a library for efficient similarity search and clustering of dense vectors. It contains algorithms that search in sets of vectors of any size, up to ones that possibly do not fit in RAM. It also contains supporting code for evaluation and parameter tuning. Faiss is written in C++ with complete wrappers for Python/numpy. Some of the most useful algorithms are implemented on the GPU. It is developed primarily at Meta's [Fundamental AI Research](https://ai.facebook.com/) group. ## News @@ -48,6 +48,7 @@ The main authors of Faiss are: - [Jeff Johnson](https://github.com/wickedfoo) implemented all of the GPU Faiss - [Lucas Hosseini](https://github.com/beauby) implemented the binary indexes and the build system - [Chengqi Deng](https://github.com/KinglittleQ) implemented NSG, NNdescent and much of the additive quantization code. +- [Alexandr Guzhva](https://github.com/alexanderguzhva) many optimizations: SIMD, memory allocation and layout, fast decoding kernels for vector codecs, etc. ## Reference diff --git a/benchs/bench_gpu_1bn.py b/benchs/bench_gpu_1bn.py index 362b10387..94e68d2fa 100644 --- a/benchs/bench_gpu_1bn.py +++ b/benchs/bench_gpu_1bn.py @@ -13,7 +13,7 @@ import sys import faiss import re -from multiprocessing.dummy import Pool as ThreadPool +from multiprocessing.pool import ThreadPool from datasets import ivecs_read #################################################################### diff --git a/benchs/bench_ivf_selector.cpp b/benchs/bench_ivf_selector.cpp index 671c6d172..6610ce1c9 100644 --- a/benchs/bench_ivf_selector.cpp +++ b/benchs/bench_ivf_selector.cpp @@ -25,7 +25,7 @@ */ int main() { - using idx_t = faiss::Index::idx_t; + using idx_t = faiss::idx_t; int d = 64; size_t nb = 1024 * 1024; size_t nq = 512 * 16; diff --git a/benchs/bench_polysemous_1bn.py b/benchs/bench_polysemous_1bn.py index 0396b56b5..eabced595 100644 --- a/benchs/bench_polysemous_1bn.py +++ b/benchs/bench_polysemous_1bn.py @@ -9,7 +9,7 @@ import time import numpy as np import re import faiss -from multiprocessing.dummy import Pool as ThreadPool +from multiprocessing.pool import ThreadPool from datasets import ivecs_read diff --git a/benchs/distributed_ondisk/distributed_kmeans.py b/benchs/distributed_ondisk/distributed_kmeans.py index d706ff350..a68359fe2 100755 --- a/benchs/distributed_ondisk/distributed_kmeans.py +++ b/benchs/distributed_ondisk/distributed_kmeans.py @@ -17,7 +17,7 @@ import numpy as np import faiss -from multiprocessing.dummy import Pool as ThreadPool +from multiprocessing.pool import ThreadPool from faiss.contrib import rpc from faiss.contrib.datasets import SyntheticDataset from faiss.contrib.vecs_io import bvecs_mmap, fvecs_mmap diff --git a/benchs/distributed_ondisk/make_index_vslice.py b/benchs/distributed_ondisk/make_index_vslice.py index ca58425b2..4d26aa202 100644 --- a/benchs/distributed_ondisk/make_index_vslice.py +++ b/benchs/distributed_ondisk/make_index_vslice.py @@ -8,7 +8,7 @@ import time import numpy as np import faiss import argparse -from multiprocessing.dummy import Pool as ThreadPool +from multiprocessing.pool import ThreadPool def ivecs_mmap(fname): a = np.memmap(fname, dtype='int32', mode='r') diff --git a/benchs/distributed_ondisk/merge_to_ondisk.py b/benchs/distributed_ondisk/merge_to_ondisk.py index 5c8f3ace9..bb5750fdf 100644 --- a/benchs/distributed_ondisk/merge_to_ondisk.py +++ b/benchs/distributed_ondisk/merge_to_ondisk.py @@ -6,7 +6,7 @@ import os import faiss import argparse -from multiprocessing.dummy import Pool as ThreadPool +from multiprocessing.pool import ThreadPool if __name__ == '__main__': diff --git a/benchs/distributed_ondisk/search_server.py b/benchs/distributed_ondisk/search_server.py index 12889962c..a7445d558 100644 --- a/benchs/distributed_ondisk/search_server.py +++ b/benchs/distributed_ondisk/search_server.py @@ -64,7 +64,7 @@ if __name__ == '__main__': # Client implementation ############################################################ -from multiprocessing.dummy import Pool as ThreadPool +from multiprocessing.pool import ThreadPool import faiss import numpy as np diff --git a/benchs/link_and_code/bench_link_and_code.py b/benchs/link_and_code/bench_link_and_code.py index 34e7022dd..ed8f86d63 100755 --- a/benchs/link_and_code/bench_link_and_code.py +++ b/benchs/link_and_code/bench_link_and_code.py @@ -8,10 +8,7 @@ import os import sys import time import numpy as np -import re import faiss -from multiprocessing.dummy import Pool as ThreadPool -import pdb import argparse import datasets from datasets import sanitize diff --git a/contrib/client_server.py b/contrib/client_server.py index 99b5b12fd..ee39798e5 100755 --- a/contrib/client_server.py +++ b/contrib/client_server.py @@ -3,7 +3,7 @@ # This source code is licensed under the MIT license found in the # LICENSE file in the root directory of this source tree. -from multiprocessing.dummy import Pool as ThreadPool +from multiprocessing.pool import ThreadPool import faiss from typing import List, Tuple diff --git a/contrib/evaluation.py b/contrib/evaluation.py index 51a4a7499..14c159db3 100644 --- a/contrib/evaluation.py +++ b/contrib/evaluation.py @@ -6,7 +6,7 @@ import numpy as np import unittest -from multiprocessing.dummy import Pool as ThreadPool +from multiprocessing.pool import ThreadPool ############################################################### # Simple functions to evaluate knn results diff --git a/demos/demo_imi_flat.cpp b/demos/demo_imi_flat.cpp index d35a4c597..7713da4a8 100644 --- a/demos/demo_imi_flat.cpp +++ b/demos/demo_imi_flat.cpp @@ -132,7 +132,7 @@ int main() { k, nq); - std::vector nns(k * nq); + std::vector nns(k * nq); std::vector dis(k * nq); index.search(nq, queries.data(), k, dis.data(), nns.data()); diff --git a/demos/demo_imi_pq.cpp b/demos/demo_imi_pq.cpp index a5af6798d..b20aefb82 100644 --- a/demos/demo_imi_pq.cpp +++ b/demos/demo_imi_pq.cpp @@ -126,7 +126,7 @@ int main() { nb); std::vector database(nb * d); - std::vector ids(nb); + std::vector ids(nb); for (size_t i = 0; i < nb; i++) { for (size_t j = 0; j < d; j++) { database[i * d + j] = distrib(rng); @@ -169,7 +169,7 @@ int main() { // - given a vector float *x, finding which k centroids are // closest to it (ie to find the nearest neighbors) can be done with // - // faiss::Index::idx_t *centroid_ids = new faiss::Index::idx_t[k]; + // faiss::idx_t *centroid_ids = new faiss::idx_t[k]; // float *distances = new float[k]; // index.quantizer->search (1, x, k, dis, centroids_ids); // @@ -184,7 +184,7 @@ int main() { k, nq); - std::vector nns(k * nq); + std::vector nns(k * nq); std::vector dis(k * nq); index.search(nq, queries.data(), k, dis.data(), nns.data()); diff --git a/demos/demo_ivfpq_indexing.cpp b/demos/demo_ivfpq_indexing.cpp index 15af713f9..fa8687aad 100644 --- a/demos/demo_ivfpq_indexing.cpp +++ b/demos/demo_ivfpq_indexing.cpp @@ -118,7 +118,7 @@ int main() { k, nq); - std::vector nns(k * nq); + std::vector nns(k * nq); std::vector dis(k * nq); index.search(nq, queries.data(), k, dis.data(), nns.data()); diff --git a/demos/demo_nndescent.cpp b/demos/demo_nndescent.cpp index 6b83256a4..34594b035 100644 --- a/demos/demo_nndescent.cpp +++ b/demos/demo_nndescent.cpp @@ -58,8 +58,8 @@ int main(void) { } int k = 5; - std::vector nns(k * nq); - std::vector gt_nns(k * nq); + std::vector nns(k * nq); + std::vector gt_nns(k * nq); std::vector dis(k * nq); auto start = high_resolution_clock::now(); diff --git a/demos/demo_residual_quantizer.cpp b/demos/demo_residual_quantizer.cpp index faaa37652..6166fc108 100644 --- a/demos/demo_residual_quantizer.cpp +++ b/demos/demo_residual_quantizer.cpp @@ -24,7 +24,7 @@ int main() { /****************************************** * Generate a test dataset ******************************************/ - using idx_t = faiss::Index::idx_t; + using idx_t = faiss::idx_t; size_t d = 128; size_t nt = 10000; size_t nb = 10000; diff --git a/demos/demo_sift1M.cpp b/demos/demo_sift1M.cpp index aa4975f77..598565f85 100644 --- a/demos/demo_sift1M.cpp +++ b/demos/demo_sift1M.cpp @@ -140,8 +140,8 @@ int main() { assert(d == d2 || !"query does not have same dimension as train set"); } - size_t k; // nb of results per query in the GT - faiss::Index::idx_t* gt; // nq * k matrix of ground-truth nearest-neighbors + size_t k; // nb of results per query in the GT + faiss::idx_t* gt; // nq * k matrix of ground-truth nearest-neighbors { printf("[%.3f s] Loading ground truth for %ld queries\n", @@ -153,7 +153,7 @@ int main() { int* gt_int = ivecs_read("sift1M/sift_groundtruth.ivecs", &k, &nq2); assert(nq2 == nq || !"incorrect nb of ground truth entries"); - gt = new faiss::Index::idx_t[k * nq]; + gt = new faiss::idx_t[k * nq]; for (int i = 0; i < k * nq; i++) { gt[i] = gt_int[i]; } @@ -219,7 +219,7 @@ int main() { nq); // output buffers - faiss::Index::idx_t* I = new faiss::Index::idx_t[nq * k]; + faiss::idx_t* I = new faiss::idx_t[nq * k]; float* D = new float[nq * k]; index->search(nq, xq, k, D, I); diff --git a/demos/demo_weighted_kmeans.cpp b/demos/demo_weighted_kmeans.cpp index 9892fd08a..f6f89fae0 100644 --- a/demos/demo_weighted_kmeans.cpp +++ b/demos/demo_weighted_kmeans.cpp @@ -155,7 +155,7 @@ int main(int argc, char** argv) { faiss::IndexFlatL2 cent_index(d); cent_index.add(nc, centroids.data()); std::vector dis(n); - std::vector idx(n); + std::vector idx(n); cent_index.search( nc * 2, ccent.data(), 1, dis.data(), idx.data()); diff --git a/faiss/AutoTune.cpp b/faiss/AutoTune.cpp index dbe632206..3cec38cb4 100644 --- a/faiss/AutoTune.cpp +++ b/faiss/AutoTune.cpp @@ -608,7 +608,7 @@ void ParameterSpace::explore( if (n_experiments == 0) { for (size_t cno = 0; cno < n_comb; cno++) { set_index_parameters(index, cno); - std::vector I(nq * crit.nnn); + std::vector I(nq * crit.nnn); std::vector D(nq * crit.nnn); double t0 = getmillisecs(); @@ -677,7 +677,7 @@ void ParameterSpace::explore( } set_index_parameters(index, cno); - std::vector I(nq * crit.nnn); + std::vector I(nq * crit.nnn); std::vector D(nq * crit.nnn); double t0 = getmillisecs(); @@ -688,7 +688,7 @@ void ParameterSpace::explore( do { if (thread_over_batches) { #pragma omp parallel for - for (Index::idx_t q0 = 0; q0 < nq; q0 += batchsize) { + for (idx_t q0 = 0; q0 < nq; q0 += batchsize) { size_t q1 = q0 + batchsize; if (q1 > nq) q1 = nq; diff --git a/faiss/AutoTune.h b/faiss/AutoTune.h index 58be60f96..08046782e 100644 --- a/faiss/AutoTune.h +++ b/faiss/AutoTune.h @@ -24,7 +24,6 @@ namespace faiss { * higher is better. */ struct AutoTuneCriterion { - typedef Index::idx_t idx_t; idx_t nq; ///< nb of queries this criterion is evaluated on idx_t nnn; ///< nb of NNs that the query should request idx_t gt_nnn; ///< nb of GT NNs required to evaluate criterion diff --git a/faiss/Clustering.cpp b/faiss/Clustering.cpp index a62982df3..0a0249f89 100644 --- a/faiss/Clustering.cpp +++ b/faiss/Clustering.cpp @@ -88,8 +88,6 @@ void Clustering::train( namespace { -using idx_t = Clustering::idx_t; - idx_t subsample_training_set( const Clustering& clus, idx_t nx, @@ -624,8 +622,6 @@ ProgressiveDimClustering::ProgressiveDimClustering( namespace { -using idx_t = Index::idx_t; - void copy_columns(idx_t n, idx_t d1, const float* src, idx_t d2, float* dest) { idx_t d = std::min(d1, d2); for (idx_t i = 0; i < n; i++) { diff --git a/faiss/Clustering.h b/faiss/Clustering.h index fe19cb715..f8008d429 100644 --- a/faiss/Clustering.h +++ b/faiss/Clustering.h @@ -61,7 +61,6 @@ struct ClusteringIterationStats { * */ struct Clustering : ClusteringParameters { - typedef Index::idx_t idx_t; size_t d; ///< dimension of the vectors size_t k; ///< nb of centroids @@ -154,7 +153,6 @@ struct ProgressiveDimIndexFactory { * https://arxiv.org/abs/1509.05195 */ struct ProgressiveDimClustering : ProgressiveDimClusteringParameters { - using idx_t = Index::idx_t; size_t d; ///< dimension of the vectors size_t k; ///< nb of centroids diff --git a/faiss/IVFlib.h b/faiss/IVFlib.h index 7f74394b8..5524b41e2 100644 --- a/faiss/IVFlib.h +++ b/faiss/IVFlib.h @@ -48,8 +48,6 @@ IndexIVF* try_extract_index_ivf(Index* index); */ void merge_into(Index* index0, Index* index1, bool shift_ids); -typedef Index::idx_t idx_t; - /* Returns the cluster the embeddings belong to. * * @param index Index, which should be an IVF index diff --git a/faiss/Index.h b/faiss/Index.h index 5225465a0..f3eb0521c 100644 --- a/faiss/Index.h +++ b/faiss/Index.h @@ -62,7 +62,6 @@ struct SearchParameters { * although the internal representation may vary. */ struct Index { - using idx_t = int64_t; ///< all indices are this type using component_t = float; using distance_t = float; diff --git a/faiss/IndexBinary.h b/faiss/IndexBinary.h index 8374a3b31..35a0aff79 100644 --- a/faiss/IndexBinary.h +++ b/faiss/IndexBinary.h @@ -32,7 +32,6 @@ struct RangeSearchResult; * vectors. */ struct IndexBinary { - using idx_t = Index::idx_t; ///< all indices are this type using component_t = uint8_t; using distance_t = int32_t; diff --git a/faiss/IndexBinaryHash.cpp b/faiss/IndexBinaryHash.cpp index e8aa29806..0e449bab7 100644 --- a/faiss/IndexBinaryHash.cpp +++ b/faiss/IndexBinaryHash.cpp @@ -108,8 +108,6 @@ struct FlipEnumerator { } }; -using idx_t = Index::idx_t; - struct RangeSearchResults { int radius; RangeQueryResult& qres; @@ -353,7 +351,7 @@ template static void verify_shortlist( const IndexBinaryFlat& index, const uint8_t* q, - const std::unordered_set& shortlist, + const std::unordered_set& shortlist, SearchResults& res) { size_t code_size = index.code_size; size_t nlist = 0, ndis = 0, n0 = 0; diff --git a/faiss/IndexBinaryIVF.cpp b/faiss/IndexBinaryIVF.cpp index b9f5f4c8c..f899ee3aa 100644 --- a/faiss/IndexBinaryIVF.cpp +++ b/faiss/IndexBinaryIVF.cpp @@ -158,7 +158,7 @@ void IndexBinaryIVF::reconstruct_n(idx_t i0, idx_t ni, uint8_t* recons) const { for (idx_t list_no = 0; list_no < nlist; list_no++) { size_t list_size = invlists->list_size(list_no); - const Index::idx_t* idlist = invlists->get_ids(list_no); + const idx_t* idlist = invlists->get_ids(list_no); for (idx_t offset = 0; offset < list_size; offset++) { idx_t id = idlist[offset]; @@ -320,8 +320,6 @@ void IndexBinaryIVF::replace_invlists(InvertedLists* il, bool own) { namespace { -using idx_t = Index::idx_t; - template struct IVFBinaryScannerL2 : BinaryInvertedListScanner { HammingComputer hc; @@ -448,7 +446,7 @@ void search_knn_hamming_heap( size_t list_size = ivf.invlists->list_size(key); InvertedLists::ScopedCodes scodes(ivf.invlists, key); std::unique_ptr sids; - const Index::idx_t* ids = nullptr; + const idx_t* ids = nullptr; if (!store_pairs) { sids.reset(new InvertedLists::ScopedIds(ivf.invlists, key)); @@ -533,7 +531,7 @@ void search_knn_hamming_count( size_t list_size = ivf.invlists->list_size(key); InvertedLists::ScopedCodes scodes(ivf.invlists, key); const uint8_t* list_vecs = scodes.get(); - const Index::idx_t* ids = + const idx_t* ids = store_pairs ? nullptr : ivf.invlists->get_ids(key); for (size_t j = 0; j < list_size; j++) { diff --git a/faiss/IndexBinaryIVF.h b/faiss/IndexBinaryIVF.h index a7d7e6605..73cfe092c 100644 --- a/faiss/IndexBinaryIVF.h +++ b/faiss/IndexBinaryIVF.h @@ -209,8 +209,6 @@ struct IndexBinaryIVF : IndexBinary { }; struct BinaryInvertedListScanner { - using idx_t = Index::idx_t; - /// from now on we handle this query. virtual void set_query(const uint8_t* query_vector) = 0; diff --git a/faiss/IndexFlat.cpp b/faiss/IndexFlat.cpp index 36b9aa9b7..907f13c19 100644 --- a/faiss/IndexFlat.cpp +++ b/faiss/IndexFlat.cpp @@ -90,7 +90,7 @@ namespace { struct FlatL2Dis : FlatCodesDistanceComputer { size_t d; - Index::idx_t nb; + idx_t nb; const float* q; const float* b; size_t ndis; @@ -121,7 +121,7 @@ struct FlatL2Dis : FlatCodesDistanceComputer { struct FlatIPDis : FlatCodesDistanceComputer { size_t d; - Index::idx_t nb; + idx_t nb; const float* q; const float* b; size_t ndis; diff --git a/faiss/IndexHNSW.cpp b/faiss/IndexHNSW.cpp index 3e159400d..a34b23913 100644 --- a/faiss/IndexHNSW.cpp +++ b/faiss/IndexHNSW.cpp @@ -58,7 +58,6 @@ int sgemm_( namespace faiss { -using idx_t = Index::idx_t; using MinimaxHeap = HNSW::MinimaxHeap; using storage_idx_t = HNSW::storage_idx_t; using NodeDistFarther = HNSW::NodeDistFarther; diff --git a/faiss/IndexHNSW.h b/faiss/IndexHNSW.h index 3f871ebee..587814925 100644 --- a/faiss/IndexHNSW.h +++ b/faiss/IndexHNSW.h @@ -22,7 +22,6 @@ namespace faiss { struct IndexHNSW; struct ReconstructFromNeighbors { - typedef Index::idx_t idx_t; typedef HNSW::storage_idx_t storage_idx_t; const IndexHNSW& index; diff --git a/faiss/IndexIDMap.cpp b/faiss/IndexIDMap.cpp index 5c109b15b..1417b91fd 100644 --- a/faiss/IndexIDMap.cpp +++ b/faiss/IndexIDMap.cpp @@ -64,7 +64,7 @@ template void IndexIDMapTemplate::add_with_ids( idx_t n, const typename IndexT::component_t* x, - const typename IndexT::idx_t* xids) { + const idx_t* xids) { index->add(n, x); for (idx_t i = 0; i < n; i++) id_map.push_back(xids[i]); @@ -77,7 +77,7 @@ void IndexIDMapTemplate::search( const typename IndexT::component_t* x, idx_t k, typename IndexT::distance_t* distances, - typename IndexT::idx_t* labels, + idx_t* labels, const SearchParameters* params) const { FAISS_THROW_IF_NOT_MSG( !params, "search params not supported for this index"); @@ -91,7 +91,7 @@ void IndexIDMapTemplate::search( template void IndexIDMapTemplate::range_search( - typename IndexT::idx_t n, + idx_t n, const typename IndexT::component_t* x, typename IndexT::distance_t radius, RangeSearchResult* result, @@ -182,7 +182,7 @@ template void IndexIDMap2Template::add_with_ids( idx_t n, const typename IndexT::component_t* x, - const typename IndexT::idx_t* xids) { + const idx_t* xids) { size_t prev_ntotal = this->ntotal; IndexIDMapTemplate::add_with_ids(n, x, xids); for (size_t i = prev_ntotal; i < this->ntotal; i++) { diff --git a/faiss/IndexIDMap.h b/faiss/IndexIDMap.h index 920426f8e..0643366c9 100644 --- a/faiss/IndexIDMap.h +++ b/faiss/IndexIDMap.h @@ -18,7 +18,6 @@ namespace faiss { /** Index that translates search results to ids */ template struct IndexIDMapTemplate : IndexT { - using idx_t = typename IndexT::idx_t; using component_t = typename IndexT::component_t; using distance_t = typename IndexT::distance_t; @@ -74,7 +73,6 @@ using IndexBinaryIDMap = IndexIDMapTemplate; * implementation via a 2-way index */ template struct IndexIDMap2Template : IndexIDMapTemplate { - using idx_t = typename IndexT::idx_t; using component_t = typename IndexT::component_t; using distance_t = typename IndexT::distance_t; diff --git a/faiss/IndexIVF.cpp b/faiss/IndexIVF.cpp index 8ca71331c..cd99371a9 100644 --- a/faiss/IndexIVF.cpp +++ b/faiss/IndexIVF.cpp @@ -131,7 +131,7 @@ size_t Level1Quantizer::coarse_code_size() const { return nbyte; } -void Level1Quantizer::encode_listno(Index::idx_t list_no, uint8_t* code) const { +void Level1Quantizer::encode_listno(idx_t list_no, uint8_t* code) const { // little endian size_t nl = nlist - 1; while (nl > 0) { @@ -141,7 +141,7 @@ void Level1Quantizer::encode_listno(Index::idx_t list_no, uint8_t* code) const { } } -Index::idx_t Level1Quantizer::decode_listno(const uint8_t* code) const { +idx_t Level1Quantizer::decode_listno(const uint8_t* code) const { size_t nl = nlist - 1; int64_t list_no = 0; int nbit = 0; @@ -522,7 +522,7 @@ void IndexIVF::search_preassigned( const uint8_t* codes = scodes.get(); std::unique_ptr sids; - const Index::idx_t* ids = nullptr; + const idx_t* ids = nullptr; if (!store_pairs) { sids.reset(new InvertedLists::ScopedIds(invlists, key)); diff --git a/faiss/IndexIVF.h b/faiss/IndexIVF.h index 94a84dd28..6d9d7cddc 100644 --- a/faiss/IndexIVF.h +++ b/faiss/IndexIVF.h @@ -54,8 +54,8 @@ struct Level1Quantizer { /// compute the number of bytes required to store list ids size_t coarse_code_size() const; - void encode_listno(Index::idx_t list_no, uint8_t* code) const; - Index::idx_t decode_listno(const uint8_t* code) const; + void encode_listno(idx_t list_no, uint8_t* code) const; + idx_t decode_listno(const uint8_t* code) const; Level1Quantizer(Index* quantizer, size_t nlist); @@ -366,8 +366,6 @@ struct RangeQueryResult; * distance_to_code and scan_codes can be called in multiple * threads */ struct InvertedListScanner { - using idx_t = Index::idx_t; - idx_t list_no = -1; ///< remember current list bool keep_max = false; ///< keep maximum instead of minimum /// store positions in invlists rather than labels diff --git a/faiss/IndexIVFAdditiveQuantizer.cpp b/faiss/IndexIVFAdditiveQuantizer.cpp index ac9ae90eb..7ba8ddb26 100644 --- a/faiss/IndexIVFAdditiveQuantizer.cpp +++ b/faiss/IndexIVFAdditiveQuantizer.cpp @@ -51,7 +51,7 @@ void IndexIVFAdditiveQuantizer::train_residual(idx_t n, const float* x) { ScopeDeleter del_x(x_in == x ? nullptr : x); if (by_residual) { - std::vector idx(n); + std::vector idx(n); quantizer->assign(n, x, idx.data()); std::vector residuals(n * d); diff --git a/faiss/IndexIVFFastScan.cpp b/faiss/IndexIVFFastScan.cpp index 6855e2228..ed7e74536 100644 --- a/faiss/IndexIVFFastScan.cpp +++ b/faiss/IndexIVFFastScan.cpp @@ -229,7 +229,6 @@ void estimators_from_tables_generic( } } -using idx_t = Index::idx_t; using namespace quantize_lut; } // anonymous namespace diff --git a/faiss/IndexIVFPQ.cpp b/faiss/IndexIVFPQ.cpp index 0b6702731..07c03761e 100644 --- a/faiss/IndexIVFPQ.cpp +++ b/faiss/IndexIVFPQ.cpp @@ -197,9 +197,9 @@ void IndexIVFPQ::add_core( static float* compute_residuals( const Index* quantizer, - Index::idx_t n, + idx_t n, const float* x, - const Index::idx_t* list_nos) { + const idx_t* list_nos) { size_t d = quantizer->d; float* residuals = new float[n * d]; // TODO: parallelize? @@ -520,8 +520,6 @@ void IndexIVFPQ::precompute_table() { namespace { -using idx_t = Index::idx_t; - #define TIC t0 = get_cycles() #define TOC get_cycles() - t0 @@ -622,7 +620,7 @@ struct QueryTables { *****************************************************/ // fields specific to list - Index::idx_t key; + idx_t key; float coarse_dis; std::vector q_code; @@ -1170,7 +1168,7 @@ struct IVFPQScannerT : QueryTables { * use_sel: store or ignore the IDSelector */ template -struct IVFPQScanner : IVFPQScannerT, +struct IVFPQScanner : IVFPQScannerT, InvertedListScanner { int precompute_mode; const IDSelector* sel; @@ -1180,9 +1178,7 @@ struct IVFPQScanner : IVFPQScannerT, bool store_pairs, int precompute_mode, const IDSelector* sel) - : IVFPQScannerT( - ivfpq, - nullptr), + : IVFPQScannerT(ivfpq, nullptr), precompute_mode(precompute_mode), sel(sel) { this->store_pairs = store_pairs; diff --git a/faiss/IndexIVFSpectralHash.cpp b/faiss/IndexIVFSpectralHash.cpp index 6963d30e9..61de5aecb 100644 --- a/faiss/IndexIVFSpectralHash.cpp +++ b/faiss/IndexIVFSpectralHash.cpp @@ -213,8 +213,6 @@ struct IVFScanner : InvertedListScanner { std::vector qcode; HammingComputer hc; - using idx_t = Index::idx_t; - IVFScanner(const IndexIVFSpectralHash* index, bool store_pairs) : index(index), nbit(index->nbit), diff --git a/faiss/IndexNNDescent.cpp b/faiss/IndexNNDescent.cpp index 59c19996a..ec7bf188c 100644 --- a/faiss/IndexNNDescent.cpp +++ b/faiss/IndexNNDescent.cpp @@ -50,7 +50,6 @@ int sgemm_( namespace faiss { -using idx_t = Index::idx_t; using storage_idx_t = NNDescent::storage_idx_t; /************************************************************** diff --git a/faiss/IndexNNDescent.h b/faiss/IndexNNDescent.h index dfd833cb4..9b2532054 100644 --- a/faiss/IndexNNDescent.h +++ b/faiss/IndexNNDescent.h @@ -25,7 +25,6 @@ struct IndexNNDescent : Index { using storage_idx_t = NNDescent::storage_idx_t; /// Faiss results are 64-bit - using idx_t = Index::idx_t; // the link strcuture NNDescent nndescent; diff --git a/faiss/IndexNSG.cpp b/faiss/IndexNSG.cpp index 79e6e226c..d2fe34401 100644 --- a/faiss/IndexNSG.cpp +++ b/faiss/IndexNSG.cpp @@ -23,7 +23,6 @@ namespace faiss { -using idx_t = Index::idx_t; using namespace nsg; /************************************************************** diff --git a/faiss/IndexPQ.cpp b/faiss/IndexPQ.cpp index a18f8b69c..7988199a2 100644 --- a/faiss/IndexPQ.cpp +++ b/faiss/IndexPQ.cpp @@ -74,7 +74,7 @@ template struct PQDistanceComputer : FlatCodesDistanceComputer { size_t d; MetricType metric; - Index::idx_t nb; + idx_t nb; const ProductQuantizer& pq; const float* sdc; std::vector precomputed_table; diff --git a/faiss/IndexRefine.cpp b/faiss/IndexRefine.cpp index d681b4148..ccb6b61ca 100644 --- a/faiss/IndexRefine.cpp +++ b/faiss/IndexRefine.cpp @@ -62,7 +62,7 @@ void IndexRefine::reset() { namespace { -typedef faiss::Index::idx_t idx_t; +typedef faiss::idx_t idx_t; template static void reorder_2_heaps( diff --git a/faiss/IndexReplicas.cpp b/faiss/IndexReplicas.cpp index 5a874b496..39a07f099 100644 --- a/faiss/IndexReplicas.cpp +++ b/faiss/IndexReplicas.cpp @@ -123,14 +123,13 @@ void IndexReplicasTemplate::search( size_t componentsPerVec = sizeof(component_t) == 1 ? (dim + 7) / 8 : dim; // Partition the query by the number of indices we have - faiss::Index::idx_t queriesPerIndex = - (faiss::Index::idx_t)(n + this->count() - 1) / - (faiss::Index::idx_t)this->count(); + faiss::idx_t queriesPerIndex = + (faiss::idx_t)(n + this->count() - 1) / (faiss::idx_t)this->count(); FAISS_ASSERT(n / queriesPerIndex <= this->count()); auto fn = [queriesPerIndex, componentsPerVec, n, x, k, distances, labels]( int i, const IndexT* index) { - faiss::Index::idx_t base = (faiss::Index::idx_t)i * queriesPerIndex; + faiss::idx_t base = (faiss::idx_t)i * queriesPerIndex; if (base < n) { auto numForIndex = std::min(queriesPerIndex, n - base); diff --git a/faiss/IndexReplicas.h b/faiss/IndexReplicas.h index bf79e8434..84b12a7b8 100644 --- a/faiss/IndexReplicas.h +++ b/faiss/IndexReplicas.h @@ -20,7 +20,6 @@ namespace faiss { template class IndexReplicasTemplate : public ThreadedIndex { public: - using idx_t = typename IndexT::idx_t; using component_t = typename IndexT::component_t; using distance_t = typename IndexT::distance_t; diff --git a/faiss/IndexRowwiseMinMax.cpp b/faiss/IndexRowwiseMinMax.cpp index de6925c39..a82063c0e 100644 --- a/faiss/IndexRowwiseMinMax.cpp +++ b/faiss/IndexRowwiseMinMax.cpp @@ -11,7 +11,7 @@ namespace faiss { namespace { -using idx_t = faiss::Index::idx_t; +using idx_t = faiss::idx_t; struct StorageMinMaxFP16 { uint16_t scaler; diff --git a/faiss/IndexShards.cpp b/faiss/IndexShards.cpp index 78ce9e7d7..fa8d11d94 100644 --- a/faiss/IndexShards.cpp +++ b/faiss/IndexShards.cpp @@ -22,8 +22,6 @@ namespace faiss { // subroutines namespace { -typedef Index::idx_t idx_t; - // add translation to all valid labels void translate_labels(long n, idx_t* labels, long translation) { if (translation == 0) diff --git a/faiss/IndexShards.h b/faiss/IndexShards.h index 33245ae22..648f545df 100644 --- a/faiss/IndexShards.h +++ b/faiss/IndexShards.h @@ -18,7 +18,6 @@ namespace faiss { */ template struct IndexShardsTemplate : public ThreadedIndex { - using idx_t = typename IndexT::idx_t; using component_t = typename IndexT::component_t; using distance_t = typename IndexT::distance_t; diff --git a/faiss/MetricType.h b/faiss/MetricType.h index e1fb8cfbf..404028145 100644 --- a/faiss/MetricType.h +++ b/faiss/MetricType.h @@ -10,6 +10,8 @@ #ifndef FAISS_METRIC_TYPE_H #define FAISS_METRIC_TYPE_H +#include + namespace faiss { /// The metric space for vector comparison for Faiss indices and algorithms. @@ -31,6 +33,9 @@ enum MetricType { METRIC_JensenShannon, }; +/// all vector indices are this type +using idx_t = int64_t; + } // namespace faiss #endif diff --git a/faiss/VectorTransform.cpp b/faiss/VectorTransform.cpp index 1cf24b7b5..252f4aa87 100644 --- a/faiss/VectorTransform.cpp +++ b/faiss/VectorTransform.cpp @@ -135,7 +135,7 @@ int dgesvd_( * VectorTransform *********************************************/ -float* VectorTransform::apply(Index::idx_t n, const float* x) const { +float* VectorTransform::apply(idx_t n, const float* x) const { float* xt = new float[n * d_out]; apply_noalloc(n, x, xt); return xt; @@ -166,8 +166,7 @@ LinearTransform::LinearTransform(int d_in, int d_out, bool have_bias) is_trained = false; // will be trained when A and b are initialized } -void LinearTransform::apply_noalloc(Index::idx_t n, const float* x, float* xt) - const { +void LinearTransform::apply_noalloc(idx_t n, const float* x, float* xt) const { FAISS_THROW_IF_NOT_MSG(is_trained, "Transformation not trained yet"); float c_factor; @@ -348,7 +347,7 @@ void RandomRotationMatrix::init(int seed) { is_trained = true; } -void RandomRotationMatrix::train(Index::idx_t /*n*/, const float* /*x*/) { +void RandomRotationMatrix::train(idx_t /*n*/, const float* /*x*/) { // initialize with some arbitrary seed init(12345); } @@ -442,7 +441,7 @@ void eig(size_t d_in, double* cov, double* eigenvalues, int verbose) { } // namespace -void PCAMatrix::train(Index::idx_t n, const float* x) { +void PCAMatrix::train(idx_t n, const float* x) { const float* x_in = x; x = fvecs_maybe_subsample( @@ -733,7 +732,7 @@ ITQMatrix::ITQMatrix(int d) : LinearTransform(d, d, false), max_iter(50), seed(123) {} /** translated from fbcode/deeplearning/catalyzer/catalyzer/quantizers.py */ -void ITQMatrix::train(Index::idx_t n, const float* xf) { +void ITQMatrix::train(idx_t n, const float* xf) { size_t d = d_in; std::vector rotation(d * d); @@ -957,8 +956,7 @@ void ITQTransform::train(idx_t n, const float* x) { is_trained = true; } -void ITQTransform::apply_noalloc(Index::idx_t n, const float* x, float* xt) - const { +void ITQTransform::apply_noalloc(idx_t n, const float* x, float* xt) const { FAISS_THROW_IF_NOT_MSG(is_trained, "Transformation not trained yet"); std::unique_ptr x_norm(new float[n * d_in]); @@ -1003,7 +1001,7 @@ OPQMatrix::OPQMatrix(int d, int M, int d2) pq = nullptr; } -void OPQMatrix::train(Index::idx_t n, const float* x) { +void OPQMatrix::train(idx_t n, const float* x) { const float* x_in = x; x = fvecs_maybe_subsample(d_in, (size_t*)&n, max_train_points, x, verbose); @@ -1261,7 +1259,7 @@ CenteringTransform::CenteringTransform(int d) : VectorTransform(d, d) { is_trained = false; } -void CenteringTransform::train(Index::idx_t n, const float* x) { +void CenteringTransform::train(idx_t n, const float* x) { FAISS_THROW_IF_NOT_MSG(n > 0, "need at least one training vector"); mean.resize(d_in, 0); for (idx_t i = 0; i < n; i++) { diff --git a/faiss/VectorTransform.h b/faiss/VectorTransform.h index ef1ccfdd8..d14f74c83 100644 --- a/faiss/VectorTransform.h +++ b/faiss/VectorTransform.h @@ -23,8 +23,6 @@ namespace faiss { /** Any transformation applied on a set of vectors */ struct VectorTransform { - typedef Index::idx_t idx_t; - int d_in; ///! input dimension int d_out; ///! output dimension diff --git a/faiss/gpu/GpuCloner.cpp b/faiss/gpu/GpuCloner.cpp index 4b242972b..74335626e 100644 --- a/faiss/gpu/GpuCloner.cpp +++ b/faiss/gpu/GpuCloner.cpp @@ -116,7 +116,6 @@ ToGpuCloner::ToGpuCloner( : GpuClonerOptions(options), provider(prov), device(device) {} Index* ToGpuCloner::clone_Index(const Index* index) { - using idx_t = Index::idx_t; if (auto ifl = dynamic_cast(index)) { GpuIndexFlatConfig config; config.device = device; diff --git a/faiss/gpu/GpuDistance.cu b/faiss/gpu/GpuDistance.cu index e80e3f293..1e867f2c8 100644 --- a/faiss/gpu/GpuDistance.cu +++ b/faiss/gpu/GpuDistance.cu @@ -137,20 +137,18 @@ void bfKnnConvert(GpuResourcesProvider* prov, const GpuDistanceParams& args) { args.ignoreOutDistances); // Convert and copy int indices out - auto tOutIndices = toDeviceTemporary( + auto tOutIndices = toDeviceTemporary( res, device, - (Index::idx_t*)args.outIndices, + (idx_t*)args.outIndices, stream, {args.numQueries, args.k}); // Convert int to idx_t - convertTensor( - stream, tOutIntIndices, tOutIndices); + convertTensor(stream, tOutIntIndices, tOutIndices); // Copy back if necessary - fromDevice( - tOutIndices, (Index::idx_t*)args.outIndices, stream); + fromDevice(tOutIndices, (idx_t*)args.outIndices, stream); } else if (args.outIndicesType == IndicesDataType::I32) { // We can use the brute-force API directly, as it takes i32 indices @@ -229,7 +227,7 @@ void bruteForceKnn( float* outDistances, // A region of memory size numQueries x k, with k // innermost - Index::idx_t* outIndices) { + idx_t* outIndices) { std::cerr << "bruteForceKnn is deprecated; call bfKnn instead" << std::endl; GpuDistanceParams args; diff --git a/faiss/gpu/GpuDistance.h b/faiss/gpu/GpuDistance.h index 4ed644c75..df0c8fa60 100644 --- a/faiss/gpu/GpuDistance.h +++ b/faiss/gpu/GpuDistance.h @@ -163,7 +163,7 @@ void bruteForceKnn( float* outDistances, // A region of memory size numQueries x k, with k // innermost (row major) - Index::idx_t* outIndices); + idx_t* outIndices); } // namespace gpu } // namespace faiss diff --git a/faiss/gpu/GpuIndex.cu b/faiss/gpu/GpuIndex.cu index 9448d04e3..718be0ef7 100644 --- a/faiss/gpu/GpuIndex.cu +++ b/faiss/gpu/GpuIndex.cu @@ -100,15 +100,12 @@ size_t GpuIndex::getMinPagingSize() const { return minPagedSize_; } -void GpuIndex::add(Index::idx_t n, const float* x) { +void GpuIndex::add(idx_t n, const float* x) { // Pass to add_with_ids add_with_ids(n, x, nullptr); } -void GpuIndex::add_with_ids( - Index::idx_t n, - const float* x, - const Index::idx_t* ids) { +void GpuIndex::add_with_ids(idx_t n, const float* x, const idx_t* ids) { DeviceScope scope(config_.device); FAISS_THROW_IF_NOT_MSG(this->is_trained, "Index not trained"); @@ -119,13 +116,13 @@ void GpuIndex::add_with_ids( return; } - std::vector generatedIds; + std::vector generatedIds; // Generate IDs if we need them if (!ids && addImplRequiresIDs_()) { - generatedIds = std::vector(n); + generatedIds = std::vector(n); - for (Index::idx_t i = 0; i < n; ++i) { + for (idx_t i = 0; i < n; ++i) { generatedIds[i] = this->ntotal + i; } } @@ -133,7 +130,7 @@ void GpuIndex::add_with_ids( addPaged_((int)n, x, ids ? ids : generatedIds.data()); } -void GpuIndex::addPaged_(int n, const float* x, const Index::idx_t* ids) { +void GpuIndex::addPaged_(int n, const float* x, const idx_t* ids) { if (n > 0) { size_t totalSize = (size_t)n * this->d * sizeof(float); @@ -162,7 +159,7 @@ void GpuIndex::addPaged_(int n, const float* x, const Index::idx_t* ids) { } } -void GpuIndex::addPage_(int n, const float* x, const Index::idx_t* ids) { +void GpuIndex::addPage_(int n, const float* x, const idx_t* ids) { // At this point, `x` can be resident on CPU or GPU, and `ids` may be // resident on CPU, GPU or may be null. // @@ -178,10 +175,10 @@ void GpuIndex::addPage_(int n, const float* x, const Index::idx_t* ids) { {n, this->d}); if (ids) { - auto indices = toDeviceTemporary( + auto indices = toDeviceTemporary( resources_.get(), config_.device, - const_cast(ids), + const_cast(ids), stream, {n}); @@ -191,11 +188,7 @@ void GpuIndex::addPage_(int n, const float* x, const Index::idx_t* ids) { } } -void GpuIndex::assign( - Index::idx_t n, - const float* x, - Index::idx_t* labels, - Index::idx_t k) const { +void GpuIndex::assign(idx_t n, const float* x, idx_t* labels, idx_t k) const { DeviceScope scope(config_.device); FAISS_THROW_IF_NOT_MSG(this->is_trained, "Index not trained"); @@ -216,11 +209,11 @@ void GpuIndex::assign( } void GpuIndex::search( - Index::idx_t n, + idx_t n, const float* x, - Index::idx_t k, + idx_t k, float* distances, - Index::idx_t* labels, + idx_t* labels, const SearchParameters* params) const { DeviceScope scope(config_.device); FAISS_THROW_IF_NOT_MSG(this->is_trained, "Index not trained"); @@ -251,7 +244,7 @@ void GpuIndex::search( stream, {(int)n, (int)k}); - auto outLabels = toDeviceTemporary( + auto outLabels = toDeviceTemporary( resources_.get(), config_.device, labels, stream, {(int)n, (int)k}); bool usePaged = false; @@ -278,7 +271,7 @@ void GpuIndex::search( // Copy back if necessary fromDevice(outDistances, distances, stream); - fromDevice(outLabels, labels, stream); + fromDevice(outLabels, labels, stream); } void GpuIndex::search_and_reconstruct( @@ -298,7 +291,7 @@ void GpuIndex::searchNonPaged_( const float* x, int k, float* outDistancesData, - Index::idx_t* outIndicesData, + idx_t* outIndicesData, const SearchParameters* params) const { auto stream = resources_->getDefaultStream(config_.device); @@ -319,10 +312,10 @@ void GpuIndex::searchFromCpuPaged_( const float* x, int k, float* outDistancesData, - Index::idx_t* outIndicesData, + idx_t* outIndicesData, const SearchParameters* params) const { Tensor outDistances(outDistancesData, {n, k}); - Tensor outIndices(outIndicesData, {n, k}); + Tensor outIndices(outIndicesData, {n, k}); // Is pinned memory available? auto pinnedAlloc = resources_->getPinnedMemory(); @@ -495,18 +488,16 @@ void GpuIndex::searchFromCpuPaged_( } } -void GpuIndex::compute_residual( - const float* x, - float* residual, - Index::idx_t key) const { +void GpuIndex::compute_residual(const float* x, float* residual, idx_t key) + const { FAISS_THROW_MSG("compute_residual not implemented for this type of index"); } void GpuIndex::compute_residual_n( - Index::idx_t n, + idx_t n, const float* xs, float* residuals, - const Index::idx_t* keys) const { + const idx_t* keys) const { FAISS_THROW_MSG( "compute_residual_n not implemented for this type of index"); } diff --git a/faiss/gpu/GpuIndex.h b/faiss/gpu/GpuIndex.h index 0f50d491f..0019ed6ae 100644 --- a/faiss/gpu/GpuIndex.h +++ b/faiss/gpu/GpuIndex.h @@ -51,30 +51,26 @@ class GpuIndex : public faiss::Index { /// `x` can be resident on the CPU or any GPU; copies are performed /// as needed /// Handles paged adds if the add set is too large; calls addInternal_ - void add(Index::idx_t, const float* x) override; + void add(idx_t, const float* x) override; /// `x` and `ids` can be resident on the CPU or any GPU; copies are /// performed as needed /// Handles paged adds if the add set is too large; calls addInternal_ - void add_with_ids(Index::idx_t n, const float* x, const Index::idx_t* ids) - override; + void add_with_ids(idx_t n, const float* x, const idx_t* ids) override; /// `x` and `labels` can be resident on the CPU or any GPU; copies are /// performed as needed - void assign( - Index::idx_t n, - const float* x, - Index::idx_t* labels, - Index::idx_t k = 1) const override; + void assign(idx_t n, const float* x, idx_t* labels, idx_t k = 1) + const override; /// `x`, `distances` and `labels` can be resident on the CPU or any /// GPU; copies are performed as needed void search( - Index::idx_t n, + idx_t n, const float* x, - Index::idx_t k, + idx_t k, float* distances, - Index::idx_t* labels, + idx_t* labels, const SearchParameters* params = nullptr) const override; /// `x`, `distances` and `labels` and `recons` can be resident on the CPU or @@ -90,16 +86,16 @@ class GpuIndex : public faiss::Index { /// Overridden to force GPU indices to provide their own GPU-friendly /// implementation - void compute_residual(const float* x, float* residual, Index::idx_t key) + void compute_residual(const float* x, float* residual, idx_t key) const override; /// Overridden to force GPU indices to provide their own GPU-friendly /// implementation void compute_residual_n( - Index::idx_t n, + idx_t n, const float* xs, float* residuals, - const Index::idx_t* keys) const override; + const idx_t* keys) const override; protected: /// Copy what we need from the CPU equivalent @@ -114,7 +110,7 @@ class GpuIndex : public faiss::Index { /// Overridden to actually perform the add /// All data is guaranteed to be resident on our device - virtual void addImpl_(int n, const float* x, const Index::idx_t* ids) = 0; + virtual void addImpl_(int n, const float* x, const idx_t* ids) = 0; /// Overridden to actually perform the search /// All data is guaranteed to be resident on our device @@ -123,16 +119,16 @@ class GpuIndex : public faiss::Index { const float* x, int k, float* distances, - Index::idx_t* labels, + idx_t* labels, const SearchParameters* params) const = 0; private: /// Handles paged adds if the add set is too large, passes to /// addImpl_ to actually perform the add for the current page - void addPaged_(int n, const float* x, const Index::idx_t* ids); + void addPaged_(int n, const float* x, const idx_t* ids); /// Calls addImpl_ for a single page of GPU-resident data - void addPage_(int n, const float* x, const Index::idx_t* ids); + void addPage_(int n, const float* x, const idx_t* ids); /// Calls searchImpl_ for a single page of GPU-resident data void searchNonPaged_( @@ -140,7 +136,7 @@ class GpuIndex : public faiss::Index { const float* x, int k, float* outDistancesData, - Index::idx_t* outIndicesData, + idx_t* outIndicesData, const SearchParameters* params) const; /// Calls searchImpl_ for a single page of GPU-resident data, @@ -150,7 +146,7 @@ class GpuIndex : public faiss::Index { const float* x, int k, float* outDistancesData, - Index::idx_t* outIndicesData, + idx_t* outIndicesData, const SearchParameters* params) const; protected: diff --git a/faiss/gpu/GpuIndexBinaryFlat.cu b/faiss/gpu/GpuIndexBinaryFlat.cu index c57dfe258..86377e492 100644 --- a/faiss/gpu/GpuIndexBinaryFlat.cu +++ b/faiss/gpu/GpuIndexBinaryFlat.cu @@ -78,7 +78,7 @@ void GpuIndexBinaryFlat::copyFrom(const faiss::IndexBinaryFlat* index) { // GPU code has 32 bit indices FAISS_THROW_IF_NOT_FMT( - index->ntotal <= (Index::idx_t)std::numeric_limits::max(), + index->ntotal <= (idx_t)std::numeric_limits::max(), "GPU index only supports up to %zu indices; " "attempting to copy CPU index with %zu parameters", (size_t)std::numeric_limits::max(), @@ -117,7 +117,7 @@ void GpuIndexBinaryFlat::copyTo(faiss::IndexBinaryFlat* index) const { } } -void GpuIndexBinaryFlat::add(faiss::IndexBinary::idx_t n, const uint8_t* x) { +void GpuIndexBinaryFlat::add(faiss::idx_t n, const uint8_t* x) { DeviceScope scope(binaryFlatConfig_.device); validateNumVectors(n); @@ -129,7 +129,7 @@ void GpuIndexBinaryFlat::add(faiss::IndexBinary::idx_t n, const uint8_t* x) { // Due to GPU indexing in int32, we can't store more than this // number of vectors on a GPU FAISS_THROW_IF_NOT_FMT( - this->ntotal + n <= (Index::idx_t)std::numeric_limits::max(), + this->ntotal + n <= (idx_t)std::numeric_limits::max(), "GPU index only supports up to %zu indices", (size_t)std::numeric_limits::max()); @@ -149,11 +149,11 @@ void GpuIndexBinaryFlat::reset() { } void GpuIndexBinaryFlat::search( - faiss::IndexBinary::idx_t n, + faiss::idx_t n, const uint8_t* x, - faiss::IndexBinary::idx_t k, + faiss::idx_t k, int32_t* distances, - faiss::IndexBinary::idx_t* labels, + faiss::idx_t* labels, const SearchParameters* params) const { DeviceScope scope(binaryFlatConfig_.device); auto stream = resources_->getDefaultStream(binaryFlatConfig_.device); @@ -209,7 +209,7 @@ void GpuIndexBinaryFlat::search( } // Convert and copy int indices out - auto outIndices = toDeviceTemporary( + auto outIndices = toDeviceTemporary( resources_.get(), binaryFlatConfig_.device, labels, @@ -217,11 +217,11 @@ void GpuIndexBinaryFlat::search( {(int)n, (int)k}); // Convert int to idx_t - convertTensor(stream, outIntIndices, outIndices); + convertTensor(stream, outIntIndices, outIndices); // Copy back if necessary fromDevice(outDistances, distances, stream); - fromDevice(outIndices, labels, stream); + fromDevice(outIndices, labels, stream); } void GpuIndexBinaryFlat::searchNonPaged_( @@ -277,9 +277,7 @@ void GpuIndexBinaryFlat::searchFromCpuPaged_( } } -void GpuIndexBinaryFlat::reconstruct( - faiss::IndexBinary::idx_t key, - uint8_t* out) const { +void GpuIndexBinaryFlat::reconstruct(faiss::idx_t key, uint8_t* out) const { DeviceScope scope(binaryFlatConfig_.device); FAISS_THROW_IF_NOT_MSG(key < this->ntotal, "index out of bounds"); diff --git a/faiss/gpu/GpuIndexBinaryFlat.h b/faiss/gpu/GpuIndexBinaryFlat.h index cb1c77bcc..fd97bebe2 100644 --- a/faiss/gpu/GpuIndexBinaryFlat.h +++ b/faiss/gpu/GpuIndexBinaryFlat.h @@ -53,20 +53,19 @@ class GpuIndexBinaryFlat : public IndexBinary { /// in the index instance void copyTo(faiss::IndexBinaryFlat* index) const; - void add(faiss::IndexBinary::idx_t n, const uint8_t* x) override; + void add(faiss::idx_t n, const uint8_t* x) override; void reset() override; void search( - faiss::IndexBinary::idx_t n, + faiss::idx_t n, const uint8_t* x, - faiss::IndexBinary::idx_t k, + faiss::idx_t k, int32_t* distances, - faiss::IndexBinary::idx_t* labels, + faiss::idx_t* labels, const faiss::SearchParameters* params = nullptr) const override; - void reconstruct(faiss::IndexBinary::idx_t key, uint8_t* recons) - const override; + void reconstruct(faiss::idx_t key, uint8_t* recons) const override; protected: /// Called from search when the input data is on the CPU; diff --git a/faiss/gpu/GpuIndexFlat.cu b/faiss/gpu/GpuIndexFlat.cu index 963713515..749df7c09 100644 --- a/faiss/gpu/GpuIndexFlat.cu +++ b/faiss/gpu/GpuIndexFlat.cu @@ -102,7 +102,7 @@ void GpuIndexFlat::copyFrom(const faiss::IndexFlat* index) { // GPU code has 32 bit indices FAISS_THROW_IF_NOT_FMT( - index->ntotal <= (Index::idx_t)std::numeric_limits::max(), + index->ntotal <= (idx_t)std::numeric_limits::max(), "GPU index only supports up to %zu indices; " "attempting to copy CPU index with %zu parameters", (size_t)std::numeric_limits::max(), @@ -153,11 +153,11 @@ void GpuIndexFlat::reset() { this->ntotal = 0; } -void GpuIndexFlat::train(Index::idx_t n, const float* x) { +void GpuIndexFlat::train(idx_t n, const float* x) { // nothing to do } -void GpuIndexFlat::add(Index::idx_t n, const float* x) { +void GpuIndexFlat::add(idx_t n, const float* x) { DeviceScope scope(config_.device); FAISS_THROW_IF_NOT_MSG(this->is_trained, "Index not trained"); @@ -186,7 +186,7 @@ bool GpuIndexFlat::addImplRequiresIDs_() const { return false; } -void GpuIndexFlat::addImpl_(int n, const float* x, const Index::idx_t* ids) { +void GpuIndexFlat::addImpl_(int n, const float* x, const idx_t* ids) { // current device already set // n already validated FAISS_ASSERT(data_); @@ -198,7 +198,7 @@ void GpuIndexFlat::addImpl_(int n, const float* x, const Index::idx_t* ids) { // Due to GPU indexing in int32, we can't store more than this // number of vectors on a GPU FAISS_THROW_IF_NOT_FMT( - this->ntotal + n <= (Index::idx_t)std::numeric_limits::max(), + this->ntotal + n <= (idx_t)std::numeric_limits::max(), "GPU index only supports up to %zu indices", (size_t)std::numeric_limits::max()); @@ -211,7 +211,7 @@ void GpuIndexFlat::searchImpl_( const float* x, int k, float* distances, - Index::idx_t* labels, + idx_t* labels, const SearchParameters* params) const { // current device already set // n/k already validated @@ -220,7 +220,7 @@ void GpuIndexFlat::searchImpl_( // Input and output data are already resident on the GPU Tensor queries(const_cast(x), {n, (int)this->d}); Tensor outDistances(distances, {n, k}); - Tensor outLabels(labels, {n, k}); + Tensor outLabels(labels, {n, k}); // FlatIndex only supports int indices DeviceTensor outIntLabels( @@ -236,10 +236,10 @@ void GpuIndexFlat::searchImpl_( true); // Convert int to idx_t - convertTensor(stream, outIntLabels, outLabels); + convertTensor(stream, outIntLabels, outLabels); } -void GpuIndexFlat::reconstruct(Index::idx_t key, float* out) const { +void GpuIndexFlat::reconstruct(idx_t key, float* out) const { DeviceScope scope(config_.device); FAISS_THROW_IF_NOT_FMT( @@ -262,8 +262,7 @@ void GpuIndexFlat::reconstruct(Index::idx_t key, float* out) const { fromDevice(vec.data(), out, this->d, stream); } -void GpuIndexFlat::reconstruct_n(Index::idx_t i0, Index::idx_t n, float* out) - const { +void GpuIndexFlat::reconstruct_n(idx_t i0, idx_t n, float* out) const { DeviceScope scope(config_.device); if (n == 0) { @@ -284,10 +283,10 @@ void GpuIndexFlat::reconstruct_n(Index::idx_t i0, Index::idx_t n, float* out) i0 + n - 1, this->ntotal); FAISS_THROW_IF_NOT_FMT( - n <= (Index::idx_t)std::numeric_limits::max(), + n <= (idx_t)std::numeric_limits::max(), "number of vectors requested (%zu) must be less than %zu", n, - (Index::idx_t)std::numeric_limits::max()); + (idx_t)std::numeric_limits::max()); auto stream = resources_->getDefaultStream(config_.device); auto outDevice = toDeviceTemporary( @@ -303,10 +302,8 @@ void GpuIndexFlat::reconstruct_n(Index::idx_t i0, Index::idx_t n, float* out) fromDevice(outDevice, out, stream); } -void GpuIndexFlat::reconstruct_batch( - Index::idx_t n, - const Index::idx_t* keys, - float* out) const { +void GpuIndexFlat::reconstruct_batch(idx_t n, const idx_t* keys, float* out) + const { DeviceScope scope(config_.device); auto stream = resources_->getDefaultStream(config_.device); @@ -317,10 +314,10 @@ void GpuIndexFlat::reconstruct_batch( validateNumVectors(n); - auto keysDevice = toDeviceTemporary( + auto keysDevice = toDeviceTemporary( resources_.get(), config_.device, - const_cast(keys), + const_cast(keys), stream, {(int)n}); @@ -338,18 +335,16 @@ void GpuIndexFlat::reconstruct_batch( fromDevice(outDevice, out, stream); } -void GpuIndexFlat::compute_residual( - const float* x, - float* residual, - Index::idx_t key) const { +void GpuIndexFlat::compute_residual(const float* x, float* residual, idx_t key) + const { compute_residual_n(1, x, residual, &key); } void GpuIndexFlat::compute_residual_n( - Index::idx_t n, + idx_t n, const float* xs, float* residuals, - const Index::idx_t* keys) const { + const idx_t* keys) const { DeviceScope scope(config_.device); auto stream = resources_->getDefaultStream(config_.device); @@ -366,10 +361,10 @@ void GpuIndexFlat::compute_residual_n( const_cast(xs), stream, {(int)n, (int)this->d}); - auto idsDevice = toDeviceTemporary( + auto idsDevice = toDeviceTemporary( resources_.get(), config_.device, - const_cast(keys), + const_cast(keys), stream, {(int)n}); auto residualDevice = toDeviceTemporary( diff --git a/faiss/gpu/GpuIndexFlat.h b/faiss/gpu/GpuIndexFlat.h index 3cc56bc90..b06195320 100644 --- a/faiss/gpu/GpuIndexFlat.h +++ b/faiss/gpu/GpuIndexFlat.h @@ -82,33 +82,32 @@ class GpuIndexFlat : public GpuIndex { void reset() override; /// This index is not trained, so this does nothing - void train(Index::idx_t n, const float* x) override; + void train(idx_t n, const float* x) override; /// Overrides to avoid excessive copies - void add(Index::idx_t, const float* x) override; + void add(idx_t, const float* x) override; /// Reconstruction methods; prefer the batch reconstruct as it will /// be more efficient - void reconstruct(Index::idx_t key, float* out) const override; + void reconstruct(idx_t key, float* out) const override; /// Batch reconstruction method - void reconstruct_n(Index::idx_t i0, Index::idx_t num, float* out) - const override; + void reconstruct_n(idx_t i0, idx_t num, float* out) const override; /// Batch reconstruction method - void reconstruct_batch(Index::idx_t n, const Index::idx_t* keys, float* out) + void reconstruct_batch(idx_t n, const idx_t* keys, float* out) const override; /// Compute residual - void compute_residual(const float* x, float* residual, Index::idx_t key) + void compute_residual(const float* x, float* residual, idx_t key) const override; /// Compute residual (batch mode) void compute_residual_n( - Index::idx_t n, + idx_t n, const float* xs, float* residuals, - const Index::idx_t* keys) const override; + const idx_t* keys) const override; /// For internal access inline FlatIndex* getGpuData() { @@ -121,7 +120,7 @@ class GpuIndexFlat : public GpuIndex { bool addImplRequiresIDs_() const override; /// Called from GpuIndex for add - void addImpl_(int n, const float* x, const Index::idx_t* ids) override; + void addImpl_(int n, const float* x, const idx_t* ids) override; /// Called from GpuIndex for search void searchImpl_( @@ -129,7 +128,7 @@ class GpuIndexFlat : public GpuIndex { const float* x, int k, float* distances, - Index::idx_t* labels, + idx_t* labels, const SearchParameters* params) const override; protected: diff --git a/faiss/gpu/GpuIndexIVF.cu b/faiss/gpu/GpuIndexIVF.cu index 251cf113b..72239b261 100644 --- a/faiss/gpu/GpuIndexIVF.cu +++ b/faiss/gpu/GpuIndexIVF.cu @@ -160,7 +160,7 @@ void GpuIndexIVF::copyFrom(const faiss::IndexIVF* index) { FAISS_ASSERT(index->nlist > 0); FAISS_THROW_IF_NOT_FMT( - index->nlist <= (Index::idx_t)std::numeric_limits::max(), + index->nlist <= (idx_t)std::numeric_limits::max(), "GPU index only supports %zu inverted lists", (size_t)std::numeric_limits::max()); nlist = index->nlist; @@ -281,21 +281,21 @@ std::vector GpuIndexIVF::getListVectorData(int listId, bool gpuFormat) return baseIndex_->getListVectorData(listId, gpuFormat); } -std::vector GpuIndexIVF::getListIndices(int listId) const { +std::vector GpuIndexIVF::getListIndices(int listId) const { DeviceScope scope(config_.device); FAISS_ASSERT(baseIndex_); return baseIndex_->getListIndices(listId); } -void GpuIndexIVF::addImpl_(int n, const float* x, const Index::idx_t* xids) { +void GpuIndexIVF::addImpl_(int n, const float* x, const idx_t* xids) { // Device is already set in GpuIndex::add FAISS_ASSERT(baseIndex_); FAISS_ASSERT(n > 0); // Data is already resident on the GPU Tensor data(const_cast(x), {n, (int)this->d}); - Tensor labels(const_cast(xids), {n}); + Tensor labels(const_cast(xids), {n}); // Not all vectors may be able to be added (some may contain NaNs etc) baseIndex_->addVectors(quantizer, data, labels); @@ -310,10 +310,10 @@ void GpuIndexIVF::searchImpl_( const float* x, int k, float* distances, - Index::idx_t* labels, + idx_t* labels, const SearchParameters* params) const { // Device was already set in GpuIndex::search - Index::idx_t use_nprobe = nprobe; + idx_t use_nprobe = nprobe; if (params) { auto ivfParams = dynamic_cast(params); if (ivfParams) { @@ -341,8 +341,7 @@ void GpuIndexIVF::searchImpl_( // Data is already resident on the GPU Tensor queries(const_cast(x), {n, (int)this->d}); Tensor outDistances(distances, {n, k}); - Tensor outLabels( - const_cast(labels), {n, k}); + Tensor outLabels(const_cast(labels), {n, k}); baseIndex_->search( quantizer, queries, use_nprobe, k, outDistances, outLabels); @@ -402,10 +401,10 @@ void GpuIndexIVF::search_preassigned( stream, {(int)n, (int)use_nprobe}); - auto assignDevice = toDeviceTemporary( + auto assignDevice = toDeviceTemporary( resources_.get(), config_.device, - const_cast(assign), + const_cast(assign), stream, {(int)n, (int)use_nprobe}); @@ -416,7 +415,7 @@ void GpuIndexIVF::search_preassigned( stream, {(int)n, (int)k}); - auto outIndicesDevice = toDeviceTemporary( + auto outIndicesDevice = toDeviceTemporary( resources_.get(), config_.device, labels, stream, {(int)n, (int)k}); baseIndex_->searchPreassigned( @@ -431,7 +430,7 @@ void GpuIndexIVF::search_preassigned( // If the output was not already on the GPU, copy it back fromDevice(outDistancesDevice, distances, stream); - fromDevice(outIndicesDevice, labels, stream); + fromDevice(outIndicesDevice, labels, stream); } bool GpuIndexIVF::addImplRequiresIDs_() const { @@ -439,7 +438,7 @@ bool GpuIndexIVF::addImplRequiresIDs_() const { return true; } -void GpuIndexIVF::trainQuantizer_(Index::idx_t n, const float* x) { +void GpuIndexIVF::trainQuantizer_(idx_t n, const float* x) { DeviceScope scope(config_.device); if (n == 0) { diff --git a/faiss/gpu/GpuIndexIVF.h b/faiss/gpu/GpuIndexIVF.h index 4a80cdcb0..929e94d6d 100644 --- a/faiss/gpu/GpuIndexIVF.h +++ b/faiss/gpu/GpuIndexIVF.h @@ -91,7 +91,7 @@ class GpuIndexIVF : public GpuIndex { /// Return the vector indices contained in a particular inverted list, for /// debugging purposes. - std::vector getListIndices(int listId) const; + std::vector getListIndices(int listId) const; /// Sets the number of list probes per query void setNumProbes(int nprobe); @@ -128,10 +128,10 @@ class GpuIndexIVF : public GpuIndex { protected: void verifyIVFSettings_() const; bool addImplRequiresIDs_() const override; - void trainQuantizer_(Index::idx_t n, const float* x); + void trainQuantizer_(idx_t n, const float* x); /// Called from GpuIndex for add/add_with_ids - void addImpl_(int n, const float* x, const Index::idx_t* ids) override; + void addImpl_(int n, const float* x, const idx_t* ids) override; /// Called from GpuIndex for search void searchImpl_( @@ -139,7 +139,7 @@ class GpuIndexIVF : public GpuIndex { const float* x, int k, float* distances, - Index::idx_t* labels, + idx_t* labels, const SearchParameters* params) const override; public: diff --git a/faiss/gpu/GpuIndexIVFFlat.cu b/faiss/gpu/GpuIndexIVFFlat.cu index 84410b8a6..96c446281 100644 --- a/faiss/gpu/GpuIndexIVFFlat.cu +++ b/faiss/gpu/GpuIndexIVFFlat.cu @@ -188,12 +188,12 @@ void GpuIndexIVFFlat::updateQuantizer() { } } -void GpuIndexIVFFlat::train(Index::idx_t n, const float* x) { +void GpuIndexIVFFlat::train(idx_t n, const float* x) { DeviceScope scope(config_.device); // For now, only support <= max int results FAISS_THROW_IF_NOT_FMT( - n <= (Index::idx_t)std::numeric_limits::max(), + n <= (idx_t)std::numeric_limits::max(), "GPU index only supports up to %d indices", std::numeric_limits::max()); diff --git a/faiss/gpu/GpuIndexIVFFlat.h b/faiss/gpu/GpuIndexIVFFlat.h index 4ab6f88ef..57fee2983 100644 --- a/faiss/gpu/GpuIndexIVFFlat.h +++ b/faiss/gpu/GpuIndexIVFFlat.h @@ -85,7 +85,7 @@ class GpuIndexIVFFlat : public GpuIndexIVF { void updateQuantizer() override; /// Trains the coarse quantizer based on the given vector data - void train(Index::idx_t n, const float* x) override; + void train(idx_t n, const float* x) override; protected: /// Our configuration options diff --git a/faiss/gpu/GpuIndexIVFPQ.cu b/faiss/gpu/GpuIndexIVFPQ.cu index 7a8a2d36d..741b878e3 100644 --- a/faiss/gpu/GpuIndexIVFPQ.cu +++ b/faiss/gpu/GpuIndexIVFPQ.cu @@ -270,7 +270,7 @@ void GpuIndexIVFPQ::updateQuantizer() { } } -void GpuIndexIVFPQ::trainResidualQuantizer_(Index::idx_t n, const float* x) { +void GpuIndexIVFPQ::trainResidualQuantizer_(idx_t n, const float* x) { // Code largely copied from faiss::IndexIVFPQ auto x_in = x; @@ -288,7 +288,7 @@ void GpuIndexIVFPQ::trainResidualQuantizer_(Index::idx_t n, const float* x) { printf("computing residuals\n"); } - std::vector assign(n); + std::vector assign(n); quantizer->assign(n, x, assign.data()); std::vector residuals(n * d); @@ -347,12 +347,12 @@ void GpuIndexIVFPQ::trainResidualQuantizer_(Index::idx_t n, const float* x) { index_->setPrecomputedCodes(quantizer, usePrecomputedTables_); } -void GpuIndexIVFPQ::train(Index::idx_t n, const float* x) { +void GpuIndexIVFPQ::train(idx_t n, const float* x) { DeviceScope scope(config_.device); // For now, only support <= max int results FAISS_THROW_IF_NOT_FMT( - n <= (Index::idx_t)std::numeric_limits::max(), + n <= (idx_t)std::numeric_limits::max(), "GPU index only supports up to %d indices", std::numeric_limits::max()); diff --git a/faiss/gpu/GpuIndexIVFPQ.h b/faiss/gpu/GpuIndexIVFPQ.h index 152c0a9c4..d9e36de20 100644 --- a/faiss/gpu/GpuIndexIVFPQ.h +++ b/faiss/gpu/GpuIndexIVFPQ.h @@ -131,7 +131,7 @@ class GpuIndexIVFPQ : public GpuIndexIVF { void updateQuantizer() override; /// Trains the coarse and product quantizer based on the given vector data - void train(Index::idx_t n, const float* x) override; + void train(idx_t n, const float* x) override; public: /// Like the CPU version, we expose a publically-visible ProductQuantizer @@ -143,7 +143,7 @@ class GpuIndexIVFPQ : public GpuIndexIVF { void verifyPQSettings_() const; /// Trains the PQ quantizer based on the given vector data - void trainResidualQuantizer_(Index::idx_t n, const float* x); + void trainResidualQuantizer_(idx_t n, const float* x); protected: /// Our configuration options that we were initialized with diff --git a/faiss/gpu/GpuIndexIVFScalarQuantizer.cu b/faiss/gpu/GpuIndexIVFScalarQuantizer.cu index 8d3daede5..bc6d137d9 100644 --- a/faiss/gpu/GpuIndexIVFScalarQuantizer.cu +++ b/faiss/gpu/GpuIndexIVFScalarQuantizer.cu @@ -197,19 +197,17 @@ void GpuIndexIVFScalarQuantizer::reset() { } } -void GpuIndexIVFScalarQuantizer::trainResiduals_( - Index::idx_t n, - const float* x) { +void GpuIndexIVFScalarQuantizer::trainResiduals_(idx_t n, const float* x) { // The input is already guaranteed to be on the CPU sq.train_residual(n, x, quantizer, by_residual, verbose); } -void GpuIndexIVFScalarQuantizer::train(Index::idx_t n, const float* x) { +void GpuIndexIVFScalarQuantizer::train(idx_t n, const float* x) { DeviceScope scope(config_.device); // For now, only support <= max int results FAISS_THROW_IF_NOT_FMT( - n <= (Index::idx_t)std::numeric_limits::max(), + n <= (idx_t)std::numeric_limits::max(), "GPU index only supports up to %d indices", std::numeric_limits::max()); diff --git a/faiss/gpu/GpuIndexIVFScalarQuantizer.h b/faiss/gpu/GpuIndexIVFScalarQuantizer.h index 8de337e8a..f19631a9d 100644 --- a/faiss/gpu/GpuIndexIVFScalarQuantizer.h +++ b/faiss/gpu/GpuIndexIVFScalarQuantizer.h @@ -89,14 +89,14 @@ class GpuIndexIVFScalarQuantizer : public GpuIndexIVF { void updateQuantizer() override; /// Trains the coarse and scalar quantizer based on the given vector data - void train(Index::idx_t n, const float* x) override; + void train(idx_t n, const float* x) override; protected: /// Validates index SQ parameters void verifySQSettings_() const; /// Called from train to handle SQ residual training - void trainResiduals_(Index::idx_t n, const float* x); + void trainResiduals_(idx_t n, const float* x); public: /// Exposed like the CPU version diff --git a/faiss/gpu/impl/FlatIndex.cu b/faiss/gpu/impl/FlatIndex.cu index ff4bf5de5..bb265bc9b 100644 --- a/faiss/gpu/impl/FlatIndex.cu +++ b/faiss/gpu/impl/FlatIndex.cu @@ -163,7 +163,7 @@ void FlatIndex::query( void FlatIndex::computeResidual( Tensor& vecs, - Tensor& ids, + Tensor& ids, Tensor& residuals) { if (useFloat16_) { runCalcResidual( @@ -183,8 +183,8 @@ void FlatIndex::computeResidual( } void FlatIndex::reconstruct( - Index::idx_t start, - Index::idx_t num, + idx_t start, + idx_t num, Tensor& vecs) { auto stream = resources_->getDefaultStreamCurrentDevice(); @@ -199,7 +199,7 @@ void FlatIndex::reconstruct( } void FlatIndex::reconstruct( - Tensor& ids, + Tensor& ids, Tensor& vecs) { auto stream = resources_->getDefaultStreamCurrentDevice(); diff --git a/faiss/gpu/impl/FlatIndex.cuh b/faiss/gpu/impl/FlatIndex.cuh index d5d1ce2c6..c3241736b 100644 --- a/faiss/gpu/impl/FlatIndex.cuh +++ b/faiss/gpu/impl/FlatIndex.cuh @@ -65,19 +65,14 @@ class FlatIndex { /// Compute residual for set of vectors void computeResidual( Tensor& vecs, - Tensor& ids, + Tensor& ids, Tensor& residuals); /// Gather vectors given the set of IDs - void reconstruct( - Tensor& ids, - Tensor& vecs); + void reconstruct(Tensor& ids, Tensor& vecs); /// Gather vectors given a range of IDs - void reconstruct( - Index::idx_t start, - Index::idx_t num, - Tensor& vecs); + void reconstruct(idx_t start, idx_t num, Tensor& vecs); /// Add vectors to ourselves; the pointer passed can be on the host /// or the device diff --git a/faiss/gpu/impl/IVFAppend.cu b/faiss/gpu/impl/IVFAppend.cu index 201b43417..0634e8216 100644 --- a/faiss/gpu/impl/IVFAppend.cu +++ b/faiss/gpu/impl/IVFAppend.cu @@ -27,7 +27,7 @@ namespace gpu { // Updates the device-size array of list start pointers for codes and indices __global__ void runUpdateListPointers( - Tensor listIds, + Tensor listIds, Tensor newListLength, Tensor newCodePointers, Tensor newIndexPointers, @@ -37,7 +37,7 @@ __global__ void runUpdateListPointers( int i = blockIdx.x * blockDim.x + threadIdx.x; if (i < listIds.getSize(0)) { - Index::idx_t listId = listIds[i]; + idx_t listId = listIds[i]; listLengths[listId] = newListLength[i]; listCodes[listId] = newCodePointers[i]; listIndices[listId] = newIndexPointers[i]; @@ -45,7 +45,7 @@ __global__ void runUpdateListPointers( } void runUpdateListPointers( - Tensor& listIds, + Tensor& listIds, Tensor& newListLength, Tensor& newCodePointers, Tensor& newIndexPointers, @@ -73,9 +73,9 @@ void runUpdateListPointers( // Appends new indices for vectors being added to the IVF indices lists __global__ void ivfIndicesAppend( - Tensor listIds, + Tensor listIds, Tensor listOffset, - Tensor indices, + Tensor indices, IndicesOptions opt, void** listIndices) { int vec = blockIdx.x * blockDim.x + threadIdx.x; @@ -84,7 +84,7 @@ __global__ void ivfIndicesAppend( return; } - Index::idx_t listId = listIds[vec]; + idx_t listId = listIds[vec]; int offset = listOffset[vec]; // Add vector could be invalid (contains NaNs etc) @@ -92,20 +92,20 @@ __global__ void ivfIndicesAppend( return; } - Index::idx_t index = indices[vec]; + idx_t index = indices[vec]; if (opt == INDICES_32_BIT) { // FIXME: there could be overflow here, but where should we check this? ((int*)listIndices[listId])[offset] = (int)index; } else if (opt == INDICES_64_BIT) { - ((Index::idx_t*)listIndices[listId])[offset] = index; + ((idx_t*)listIndices[listId])[offset] = index; } } void runIVFIndicesAppend( - Tensor& listIds, + Tensor& listIds, Tensor& listOffset, - Tensor& indices, + Tensor& indices, IndicesOptions opt, DeviceVector& listIndices, cudaStream_t stream) { @@ -131,14 +131,14 @@ void runIVFIndicesAppend( template __global__ void ivfFlatAppend( - Tensor listIds, + Tensor listIds, Tensor listOffset, Tensor vecs, void** listData, Codec codec) { int vec = blockIdx.x; - Index::idx_t listId = listIds[vec]; + idx_t listId = listIds[vec]; int offset = listOffset[vec]; // Add vector could be invalid (contains NaNs etc) @@ -188,7 +188,7 @@ __global__ void ivfFlatAppend( } void runIVFFlatAppend( - Tensor& listIds, + Tensor& listIds, Tensor& listOffset, Tensor& vecs, GpuScalarQuantizer* scalarQ, @@ -261,7 +261,7 @@ void runIVFFlatAppend( } __global__ void ivfpqAppend( - Tensor listIds, + Tensor listIds, Tensor listOffset, Tensor encodings, void** listCodes) { @@ -271,7 +271,7 @@ __global__ void ivfpqAppend( return; } - Index::idx_t listId = listIds[encodingToAdd]; + idx_t listId = listIds[encodingToAdd]; int vectorNumInList = listOffset[encodingToAdd]; // Add vector could be invalid (contains NaNs etc) @@ -292,7 +292,7 @@ __global__ void ivfpqAppend( } void runIVFPQAppend( - Tensor& listIds, + Tensor& listIds, Tensor& listOffset, Tensor& encodings, DeviceVector& listCodes, @@ -344,7 +344,7 @@ template __global__ void ivfInterleavedAppend( // the IDs (offset in listData) of the unique lists // being added to - Tensor uniqueLists, + Tensor uniqueLists, // For each of the list IDs in uniqueLists, the start // offset in vectorsByUniqueList for the vectors that // we are adding to that list @@ -369,7 +369,7 @@ __global__ void ivfInterleavedAppend( int warpsPerBlock = blockDim.x / kWarpSize; // Each block is dedicated to a separate list - Index::idx_t listId = uniqueLists[blockIdx.x]; + idx_t listId = uniqueLists[blockIdx.x]; // The vecs we add to the list are at indices [vBUL[vecIdStart], // vBUL[vecIdEnd]) @@ -448,9 +448,9 @@ __global__ void ivfInterleavedAppend( } void runIVFFlatInterleavedAppend( - Tensor& listIds, + Tensor& listIds, Tensor& listOffset, - Tensor& uniqueLists, + Tensor& uniqueLists, Tensor& vectorsByUniqueList, Tensor& uniqueListVectorStart, Tensor& uniqueListStartOffset, @@ -582,9 +582,9 @@ void runIVFFlatInterleavedAppend( } void runIVFPQInterleavedAppend( - Tensor& listIds, + Tensor& listIds, Tensor& listOffset, - Tensor& uniqueLists, + Tensor& uniqueLists, Tensor& vectorsByUniqueList, Tensor& uniqueListVectorStart, Tensor& uniqueListStartOffset, diff --git a/faiss/gpu/impl/IVFAppend.cuh b/faiss/gpu/impl/IVFAppend.cuh index 38e283e83..a6c7095c3 100644 --- a/faiss/gpu/impl/IVFAppend.cuh +++ b/faiss/gpu/impl/IVFAppend.cuh @@ -18,16 +18,16 @@ namespace gpu { /// Append user indices to IVF lists void runIVFIndicesAppend( - Tensor& listIds, + Tensor& listIds, Tensor& listOffset, - Tensor& indices, + Tensor& indices, IndicesOptions opt, DeviceVector& listIndices, cudaStream_t stream); /// Update device-side list pointers in a batch void runUpdateListPointers( - Tensor& listIds, + Tensor& listIds, Tensor& newListLength, Tensor& newCodePointers, Tensor& newIndexPointers, @@ -38,7 +38,7 @@ void runUpdateListPointers( /// Append PQ codes to IVF lists (non-interleaved format) void runIVFPQAppend( - Tensor& listIds, + Tensor& listIds, Tensor& listOffset, Tensor& encodings, DeviceVector& listCodes, @@ -46,9 +46,9 @@ void runIVFPQAppend( /// Append PQ codes to IVF lists (interleaved format) void runIVFPQInterleavedAppend( - Tensor& listIds, + Tensor& listIds, Tensor& listOffset, - Tensor& uniqueLists, + Tensor& uniqueLists, Tensor& vectorsByUniqueList, Tensor& uniqueListVectorStart, Tensor& uniqueListStartOffset, @@ -59,7 +59,7 @@ void runIVFPQInterleavedAppend( /// Append SQ codes to IVF lists (non-interleaved, old format) void runIVFFlatAppend( - Tensor& listIds, + Tensor& listIds, Tensor& listOffset, Tensor& vecs, GpuScalarQuantizer* scalarQ, @@ -68,9 +68,9 @@ void runIVFFlatAppend( /// Append SQ codes to IVF lists (interleaved) void runIVFFlatInterleavedAppend( - Tensor& listIds, + Tensor& listIds, Tensor& listOffset, - Tensor& uniqueLists, + Tensor& uniqueLists, Tensor& vectorsByUniqueList, Tensor& uniqueListVectorStart, Tensor& uniqueListStartOffset, diff --git a/faiss/gpu/impl/IVFBase.cu b/faiss/gpu/impl/IVFBase.cu index bb985c22f..b4af6a40f 100644 --- a/faiss/gpu/impl/IVFBase.cu +++ b/faiss/gpu/impl/IVFBase.cu @@ -93,7 +93,7 @@ void IVFBase::reserveMemory(size_t numVecs) { // Reserve for index lists as well size_t bytesPerIndexList = vecsPerList * (indicesOptions_ == INDICES_32_BIT ? sizeof(int) - : sizeof(Index::idx_t)); + : sizeof(idx_t)); for (auto& list : deviceListIndices_) { list->data.reserve(bytesPerIndexList, stream); @@ -125,7 +125,7 @@ void IVFBase::reset() { deviceListIndices_.emplace_back(std::unique_ptr( new DeviceIVFList(resources_, info))); - listOffsetToUserIndex_.emplace_back(std::vector()); + listOffsetToUserIndex_.emplace_back(std::vector()); } deviceListDataPointers_.resize(numLists_, stream); @@ -176,7 +176,7 @@ size_t IVFBase::reclaimMemory_(bool exact) { } void IVFBase::updateDeviceListInfo_(cudaStream_t stream) { - std::vector listIds(deviceListData_.size()); + std::vector listIds(deviceListData_.size()); for (int i = 0; i < deviceListData_.size(); ++i) { listIds[i] = i; } @@ -185,9 +185,9 @@ void IVFBase::updateDeviceListInfo_(cudaStream_t stream) { } void IVFBase::updateDeviceListInfo_( - const std::vector& listIds, + const std::vector& listIds, cudaStream_t stream) { - HostTensor hostListsToUpdate({(int)listIds.size()}); + HostTensor hostListsToUpdate({(int)listIds.size()}); HostTensor hostNewListLength({(int)listIds.size()}); HostTensor hostNewDataPointers({(int)listIds.size()}); HostTensor hostNewIndexPointers({(int)listIds.size()}); @@ -204,7 +204,7 @@ void IVFBase::updateDeviceListInfo_( } // Copy the above update sets to the GPU - DeviceTensor listsToUpdate( + DeviceTensor listsToUpdate( resources_, makeTempAlloc(AllocType::Other, stream), hostListsToUpdate); @@ -250,7 +250,7 @@ int IVFBase::getListLength(int listId) const { return deviceListData_[listId]->numVecs; } -std::vector IVFBase::getListIndices(int listId) const { +std::vector IVFBase::getListIndices(int listId) const { FAISS_THROW_IF_NOT_FMT( listId < numLists_, "IVF list %d is out of bounds (%d lists total)", @@ -267,9 +267,9 @@ std::vector IVFBase::getListIndices(int listId) const { auto intInd = deviceListIndices_[listId]->data.copyToHost(stream); - std::vector out(intInd.size()); + std::vector out(intInd.size()); for (size_t i = 0; i < intInd.size(); ++i) { - out[i] = (Index::idx_t)intInd[i]; + out[i] = (idx_t)intInd[i]; } return out; @@ -277,8 +277,7 @@ std::vector IVFBase::getListIndices(int listId) const { // The data is stored as int64 on the GPU FAISS_ASSERT(listId < deviceListIndices_.size()); - return deviceListIndices_[listId]->data.copyToHost( - stream); + return deviceListIndices_[listId]->data.copyToHost(stream); } else if (indicesOptions_ == INDICES_CPU) { // The data is not stored on the GPU FAISS_ASSERT(listId < listOffsetToUserIndex_.size()); @@ -294,7 +293,7 @@ std::vector IVFBase::getListIndices(int listId) const { } else { // unhandled indices type (includes INDICES_IVF) FAISS_ASSERT(false); - return std::vector(); + return std::vector(); } } @@ -353,7 +352,7 @@ void IVFBase::copyInvertedListsTo(InvertedLists* ivf) { void IVFBase::addEncodedVectorsToList_( int listId, const void* codes, - const Index::idx_t* indices, + const idx_t* indices, size_t numVecs) { auto stream = resources_->getDefaultStreamCurrentDevice(); @@ -403,7 +402,7 @@ void IVFBase::addEncodedVectorsToList_( void IVFBase::addIndicesFromCpu_( int listId, - const Index::idx_t* indices, + const idx_t* indices, size_t numVecs) { auto stream = resources_->getDefaultStreamCurrentDevice(); @@ -417,7 +416,7 @@ void IVFBase::addIndicesFromCpu_( std::vector indices32(numVecs); for (size_t i = 0; i < numVecs; ++i) { auto ind = indices[i]; - FAISS_ASSERT(ind <= (Index::idx_t)std::numeric_limits::max()); + FAISS_ASSERT(ind <= (idx_t)std::numeric_limits::max()); indices32[i] = (int)ind; } @@ -435,7 +434,7 @@ void IVFBase::addIndicesFromCpu_( } else if (indicesOptions_ == INDICES_64_BIT) { listIndices->data.append( (uint8_t*)indices, - numVecs * sizeof(Index::idx_t), + numVecs * sizeof(idx_t), stream, true /* exact reserved size */); @@ -519,7 +518,7 @@ void IVFBase::searchCoarseQuantizer_( // Guaranteed to be on device Tensor& vecs, Tensor& distances, - Tensor& indices, + Tensor& indices, Tensor* residuals, Tensor* centroids) { auto stream = resources_->getDefaultStreamCurrentDevice(); @@ -556,7 +555,7 @@ void IVFBase::searchCoarseQuantizer_( auto cpuVecs = toHost( vecs.data(), stream, {vecs.getSize(0), vecs.getSize(1)}); auto cpuDistances = std::vector(vecs.getSize(0) * nprobe); - auto cpuIndices = std::vector(vecs.getSize(0) * nprobe); + auto cpuIndices = std::vector(vecs.getSize(0) * nprobe); coarseQuantizer->search( vecs.getSize(0), @@ -602,7 +601,7 @@ void IVFBase::searchCoarseQuantizer_( int IVFBase::addVectors( Index* coarseQuantizer, Tensor& vecs, - Tensor& indices) { + Tensor& indices) { FAISS_ASSERT(vecs.getSize(0) == indices.getSize(0)); FAISS_ASSERT(vecs.getSize(1) == dim_); @@ -617,7 +616,7 @@ int IVFBase::addVectors( {vecs.getSize(0), 1}); // We do need the closest IVF cell IDs though - DeviceTensor ivfIndices( + DeviceTensor ivfIndices( resources_, makeTempAlloc(AllocType::Other, stream), {vecs.getSize(0), 1}); @@ -648,10 +647,10 @@ int IVFBase::addVectors( // encoded vectors and indices // list id -> vectors being added - std::unordered_map> listToVectorIds; + std::unordered_map> listToVectorIds; // vector id -> which list it is being appended to - std::vector vectorIdToList(vecs.getSize(0)); + std::vector vectorIdToList(vecs.getSize(0)); // vector id -> offset in list // (we already have vector id -> list id in listIds) @@ -694,7 +693,7 @@ int IVFBase::addVectors( } // unique lists being added to - std::vector uniqueLists; + std::vector uniqueLists; for (auto& vecs : listToVectorIds) { uniqueLists.push_back(vecs.first); @@ -759,7 +758,7 @@ int IVFBase::addVectors( (indicesOptions_ == INDICES_64_BIT)) { size_t indexSize = (indicesOptions_ == INDICES_32_BIT) ? sizeof(int) - : sizeof(Index::idx_t); + : sizeof(idx_t); indices->data.resize( indices->data.size() + numVecsToAdd * indexSize, @@ -792,10 +791,10 @@ int IVFBase::addVectors( // map. We already resized our map above. if (indicesOptions_ == INDICES_CPU) { // We need to maintain the indices on the CPU side - HostTensor hostIndices(indices, stream); + HostTensor hostIndices(indices, stream); for (int i = 0; i < hostIndices.getSize(0); ++i) { - Index::idx_t listId = ivfIndicesHost[i]; + idx_t listId = ivfIndicesHost[i]; // Add vector could be invalid (contains NaNs etc) if (listId < 0) { diff --git a/faiss/gpu/impl/IVFBase.cuh b/faiss/gpu/impl/IVFBase.cuh index 5fadfe7f7..5feb577a2 100644 --- a/faiss/gpu/impl/IVFBase.cuh +++ b/faiss/gpu/impl/IVFBase.cuh @@ -62,7 +62,7 @@ class IVFBase { int getListLength(int listId) const; /// Return the list indices of a particular list back to the CPU - std::vector getListIndices(int listId) const; + std::vector getListIndices(int listId) const; /// Return the encoded vectors of a particular list back to the CPU std::vector getListVectorData(int listId, bool gpuFormat) const; @@ -84,7 +84,7 @@ class IVFBase { int addVectors( Index* coarseQuantizer, Tensor& vecs, - Tensor& indices); + Tensor& indices); /// Find the approximate k nearest neigbors for `queries` against /// our database @@ -94,7 +94,7 @@ class IVFBase { int nprobe, int k, Tensor& outDistances, - Tensor& outIndices) = 0; + Tensor& outIndices) = 0; /// Performs search when we are already given the IVF cells to look at /// (GpuIndexIVF::search_preassigned implementation) @@ -102,10 +102,10 @@ class IVFBase { Index* coarseQuantizer, Tensor& vecs, Tensor& ivfDistances, - Tensor& ivfAssignments, + Tensor& ivfAssignments, int k, Tensor& outDistances, - Tensor& outIndices, + Tensor& outIndices, bool storePairs) = 0; protected: @@ -116,7 +116,7 @@ class IVFBase { // resident on the host const void* codes, // resident on the host - const Index::idx_t* indices, + const idx_t* indices, size_t numVecs); /// Performs search in a CPU or GPU coarse quantizer for IVF cells, @@ -132,7 +132,7 @@ class IVFBase { Tensor& distances, // Output: the closest nprobe IVF cells the query vectors lie in // size (#vecs, nprobe) - Tensor& indices, + Tensor& indices, // optionally compute the residual relative to the IVF cell centroid // if passed // size (#vecs, nprobe, dim) @@ -163,12 +163,12 @@ class IVFBase { virtual void appendVectors_( Tensor& vecs, Tensor& ivfCentroidResiduals, - Tensor& indices, - Tensor& uniqueLists, + Tensor& indices, + Tensor& uniqueLists, Tensor& vectorsByUniqueList, Tensor& uniqueListVectorStart, Tensor& uniqueListStartOffset, - Tensor& listIds, + Tensor& listIds, Tensor& listOffset, cudaStream_t stream) = 0; @@ -182,14 +182,11 @@ class IVFBase { /// For a set of list IDs, update device-side list pointer and size /// information void updateDeviceListInfo_( - const std::vector& listIds, + const std::vector& listIds, cudaStream_t stream); /// Shared function to copy indices from CPU to GPU - void addIndicesFromCpu_( - int listId, - const Index::idx_t* indices, - size_t numVecs); + void addIndicesFromCpu_(int listId, const idx_t* indices, size_t numVecs); protected: /// Collection of GPU resources that we use @@ -266,7 +263,7 @@ class IVFBase { /// If we are storing indices on the CPU (indicesOptions_ is /// INDICES_CPU), then this maintains a CPU-side map of what /// (inverted list id, offset) maps to which user index - std::vector> listOffsetToUserIndex_; + std::vector> listOffsetToUserIndex_; }; } // namespace gpu diff --git a/faiss/gpu/impl/IVFFlat.cu b/faiss/gpu/impl/IVFFlat.cu index 9d7637d3d..bffa884ec 100644 --- a/faiss/gpu/impl/IVFFlat.cu +++ b/faiss/gpu/impl/IVFFlat.cu @@ -115,12 +115,12 @@ std::vector IVFFlat::translateCodesFromGpu_( void IVFFlat::appendVectors_( Tensor& vecs, Tensor& ivfCentroidResiduals, - Tensor& indices, - Tensor& uniqueLists, + Tensor& indices, + Tensor& uniqueLists, Tensor& vectorsByUniqueList, Tensor& uniqueListVectorStart, Tensor& uniqueListStartOffset, - Tensor& listIds, + Tensor& listIds, Tensor& listOffset, cudaStream_t stream) { // @@ -167,7 +167,7 @@ void IVFFlat::search( int nprobe, int k, Tensor& outDistances, - Tensor& outIndices) { + Tensor& outIndices) { auto stream = resources_->getDefaultStreamCurrentDevice(); // These are caught at a higher level @@ -185,7 +185,7 @@ void IVFFlat::search( resources_, makeTempAlloc(AllocType::Other, stream), {queries.getSize(0), nprobe}); - DeviceTensor coarseIndices( + DeviceTensor coarseIndices( resources_, makeTempAlloc(AllocType::Other, stream), {queries.getSize(0), nprobe}); @@ -223,10 +223,10 @@ void IVFFlat::searchPreassigned( Index* coarseQuantizer, Tensor& vecs, Tensor& ivfDistances, - Tensor& ivfAssignments, + Tensor& ivfAssignments, int k, Tensor& outDistances, - Tensor& outIndices, + Tensor& outIndices, bool storePairs) { FAISS_ASSERT(ivfDistances.getSize(0) == vecs.getSize(0)); FAISS_ASSERT(ivfAssignments.getSize(0) == vecs.getSize(0)); @@ -286,11 +286,11 @@ void IVFFlat::searchPreassigned( void IVFFlat::searchImpl_( Tensor& queries, Tensor& coarseDistances, - Tensor& coarseIndices, + Tensor& coarseIndices, Tensor& ivfCentroids, int k, Tensor& outDistances, - Tensor& outIndices, + Tensor& outIndices, bool storePairs) { FAISS_ASSERT(storePairs == false); @@ -336,7 +336,7 @@ void IVFFlat::searchImpl_( // FIXME: we might ultimately be calling this function with inputs // from the CPU, these are unnecessary copies if (indicesOptions_ == INDICES_CPU) { - HostTensor hostOutIndices(outIndices, stream); + HostTensor hostOutIndices(outIndices, stream); ivfOffsetToUserIndex( hostOutIndices.data(), diff --git a/faiss/gpu/impl/IVFFlat.cuh b/faiss/gpu/impl/IVFFlat.cuh index 27e931be4..1a1315c4f 100644 --- a/faiss/gpu/impl/IVFFlat.cuh +++ b/faiss/gpu/impl/IVFFlat.cuh @@ -37,7 +37,7 @@ class IVFFlat : public IVFBase { int nprobe, int k, Tensor& outDistances, - Tensor& outIndices) override; + Tensor& outIndices) override; /// Performs search when we are already given the IVF cells to look at /// (GpuIndexIVF::search_preassigned implementation) @@ -45,10 +45,10 @@ class IVFFlat : public IVFBase { Index* coarseQuantizer, Tensor& vecs, Tensor& ivfDistances, - Tensor& ivfAssignments, + Tensor& ivfAssignments, int k, Tensor& outDistances, - Tensor& outIndices, + Tensor& outIndices, bool storePairs) override; protected: @@ -73,12 +73,12 @@ class IVFFlat : public IVFBase { void appendVectors_( Tensor& vecs, Tensor& ivfCentroidResiduals, - Tensor& indices, - Tensor& uniqueLists, + Tensor& indices, + Tensor& uniqueLists, Tensor& vectorsByUniqueList, Tensor& uniqueListVectorStart, Tensor& uniqueListStartOffset, - Tensor& listIds, + Tensor& listIds, Tensor& listOffset, cudaStream_t stream) override; @@ -87,11 +87,11 @@ class IVFFlat : public IVFBase { void searchImpl_( Tensor& queries, Tensor& coarseDistances, - Tensor& coarseIndices, + Tensor& coarseIndices, Tensor& ivfCentroids, int k, Tensor& outDistances, - Tensor& outIndices, + Tensor& outIndices, bool storePairs); protected: diff --git a/faiss/gpu/impl/IVFFlatScan.cu b/faiss/gpu/impl/IVFFlatScan.cu index 19f4c3f11..09c7bc675 100644 --- a/faiss/gpu/impl/IVFFlatScan.cu +++ b/faiss/gpu/impl/IVFFlatScan.cu @@ -137,7 +137,7 @@ __global__ void ivfFlatScan( Tensor queries, bool useResidual, Tensor residualBase, - Tensor listIds, + Tensor listIds, void** allListData, int* listLengths, Codec codec, @@ -153,7 +153,7 @@ __global__ void ivfFlatScan( // We ensure that before the array (at offset -1), there is a 0 value int outBase = *(prefixSumOffsets[queryId][probeId].data() - 1); - Index::idx_t listId = listIds[queryId][probeId]; + idx_t listId = listIds[queryId][probeId]; // Safety guard in case NaNs in input cause no list ID to be generated if (listId == -1) { return; @@ -185,7 +185,7 @@ __global__ void ivfFlatScan( void runIVFFlatScanTile( GpuResources* res, Tensor& queries, - Tensor& listIds, + Tensor& listIds, DeviceVector& listData, DeviceVector& listIndices, IndicesOptions indicesOptions, @@ -201,7 +201,7 @@ void runIVFFlatScanTile( Tensor& residualBase, GpuScalarQuantizer* scalarQ, Tensor& outDistances, - Tensor& outIndices, + Tensor& outIndices, cudaStream_t stream) { int dim = queries.getSize(1); @@ -341,7 +341,7 @@ void runIVFFlatScanTile( void runIVFFlatScan( Tensor& queries, - Tensor& listIds, + Tensor& listIds, DeviceVector& listData, DeviceVector& listIndices, IndicesOptions indicesOptions, @@ -355,7 +355,7 @@ void runIVFFlatScan( // output Tensor& outDistances, // output - Tensor& outIndices, + Tensor& outIndices, GpuResources* res) { constexpr int kMinQueryTileSize = 8; constexpr int kMaxQueryTileSize = 65536; // used as blockIdx.y dimension diff --git a/faiss/gpu/impl/IVFFlatScan.cuh b/faiss/gpu/impl/IVFFlatScan.cuh index 3b61c5da3..3a5edc867 100644 --- a/faiss/gpu/impl/IVFFlatScan.cuh +++ b/faiss/gpu/impl/IVFFlatScan.cuh @@ -21,7 +21,7 @@ class GpuResources; void runIVFFlatScan( Tensor& queries, - Tensor& listIds, + Tensor& listIds, DeviceVector& listData, DeviceVector& listIndices, IndicesOptions indicesOptions, @@ -35,7 +35,7 @@ void runIVFFlatScan( // output Tensor& outDistances, // output - Tensor& outIndices, + Tensor& outIndices, GpuResources* res); } // namespace gpu diff --git a/faiss/gpu/impl/IVFInterleaved.cu b/faiss/gpu/impl/IVFInterleaved.cu index 2144b6577..c4ea9d34b 100644 --- a/faiss/gpu/impl/IVFInterleaved.cu +++ b/faiss/gpu/impl/IVFInterleaved.cu @@ -19,13 +19,13 @@ template __global__ void ivfInterleavedScan2( Tensor distanceIn, Tensor indicesIn, - Tensor listIds, + Tensor listIds, int k, void** listIndices, IndicesOptions opt, bool dir, Tensor distanceOut, - Tensor indicesOut) { + Tensor indicesOut) { int queryId = blockIdx.x; constexpr int kNumWarps = ThreadsPerBlock / kWarpSize; @@ -66,7 +66,7 @@ __global__ void ivfInterleavedScan2( uint32_t curK = i % k; uint32_t index = (curProbe << 16) | (curK & (uint32_t)0xffff); - Index::idx_t listId = listIds[queryId][curProbe]; + idx_t listId = listIds[queryId][curProbe]; if (listId != -1) { // Adjust the value we are selecting based on the sorting order heap.addThreadQ(distanceBase[i] * adj, index); @@ -81,7 +81,7 @@ __global__ void ivfInterleavedScan2( uint32_t curK = i % k; uint32_t index = (curProbe << 16) | (curK & (uint32_t)0xffff); - Index::idx_t listId = listIds[queryId][curProbe]; + idx_t listId = listIds[queryId][curProbe]; if (listId != -1) { heap.addThreadQ(distanceBase[i] * adj, index); } @@ -96,7 +96,7 @@ __global__ void ivfInterleavedScan2( auto packedIndex = smemV[i]; // We need to remap to the user-provided indices - Index::idx_t index = -1; + idx_t index = -1; // We may not have at least k values to return; in this function, max // uint32 is our sentinel value @@ -104,15 +104,15 @@ __global__ void ivfInterleavedScan2( uint32_t curProbe = packedIndex >> 16; uint32_t curK = packedIndex & 0xffff; - Index::idx_t listId = listIds[queryId][curProbe]; + idx_t listId = listIds[queryId][curProbe]; int listOffset = indicesIn[queryId][curProbe][curK]; if (opt == INDICES_32_BIT) { - index = (Index::idx_t)((int*)listIndices[listId])[listOffset]; + index = (idx_t)((int*)listIndices[listId])[listOffset]; } else if (opt == INDICES_64_BIT) { - index = ((Index::idx_t*)listIndices[listId])[listOffset]; + index = ((idx_t*)listIndices[listId])[listOffset]; } else { - index = (listId << 32 | (Index::idx_t)listOffset); + index = (listId << 32 | (idx_t)listOffset); } } @@ -123,13 +123,13 @@ __global__ void ivfInterleavedScan2( void runIVFInterleavedScan2( Tensor& distanceIn, Tensor& indicesIn, - Tensor& listIds, + Tensor& listIds, int k, DeviceVector& listIndices, IndicesOptions indicesOptions, bool dir, Tensor& distanceOut, - Tensor& indicesOut, + Tensor& indicesOut, cudaStream_t stream) { #define IVF_SCAN_2(THREADS, NUM_WARP_Q, NUM_THREAD_Q) \ ivfInterleavedScan2 \ @@ -168,7 +168,7 @@ void runIVFInterleavedScan2( void runIVFInterleavedScan( Tensor& queries, - Tensor& listIds, + Tensor& listIds, DeviceVector& listData, DeviceVector& listIndices, IndicesOptions indicesOptions, @@ -181,7 +181,7 @@ void runIVFInterleavedScan( // output Tensor& outDistances, // output - Tensor& outIndices, + Tensor& outIndices, GpuResources* res) { // caught for exceptions at a higher level FAISS_ASSERT(k <= GPU_MAX_SELECTION_K); diff --git a/faiss/gpu/impl/IVFInterleaved.cuh b/faiss/gpu/impl/IVFInterleaved.cuh index 928fe1b4a..8fd5189b4 100644 --- a/faiss/gpu/impl/IVFInterleaved.cuh +++ b/faiss/gpu/impl/IVFInterleaved.cuh @@ -40,7 +40,7 @@ template < __global__ void ivfInterleavedScan( Tensor queries, Tensor residualBase, - Tensor listIds, + Tensor listIds, void** allListData, int* listLengths, Codec codec, @@ -56,7 +56,7 @@ __global__ void ivfInterleavedScan( for (int queryId = blockIdx.y; queryId < queries.getSize(0); queryId += gridDim.y) { int probeId = blockIdx.x; - Index::idx_t listId = listIds[queryId][probeId]; + idx_t listId = listIds[queryId][probeId]; // Safety guard in case NaNs in input cause no list ID to be generated, // or we have more nprobe than nlist @@ -412,7 +412,7 @@ __global__ void ivfInterleavedScan( // with all implementations void runIVFInterleavedScan( Tensor& queries, - Tensor& listIds, + Tensor& listIds, DeviceVector& listData, DeviceVector& listIndices, IndicesOptions indicesOptions, @@ -425,7 +425,7 @@ void runIVFInterleavedScan( // output Tensor& outDistances, // output - Tensor& outIndices, + Tensor& outIndices, GpuResources* res); // Second pass of IVF list scanning to perform final k-selection and look up the @@ -433,13 +433,13 @@ void runIVFInterleavedScan( void runIVFInterleavedScan2( Tensor& distanceIn, Tensor& indicesIn, - Tensor& listIds, + Tensor& listIds, int k, DeviceVector& listIndices, IndicesOptions indicesOptions, bool dir, Tensor& distanceOut, - Tensor& indicesOut, + Tensor& indicesOut, cudaStream_t stream); } // namespace gpu diff --git a/faiss/gpu/impl/IVFPQ.cu b/faiss/gpu/impl/IVFPQ.cu index a8715045e..1cc497a61 100644 --- a/faiss/gpu/impl/IVFPQ.cu +++ b/faiss/gpu/impl/IVFPQ.cu @@ -129,12 +129,12 @@ Tensor IVFPQ::getPQCentroids() { void IVFPQ::appendVectors_( Tensor& vecs, Tensor& ivfCentroidResiduals, - Tensor& indices, - Tensor& uniqueLists, + Tensor& indices, + Tensor& uniqueLists, Tensor& vectorsByUniqueList, Tensor& uniqueListVectorStart, Tensor& uniqueListStartOffset, - Tensor& listIds, + Tensor& listIds, Tensor& listOffset, cudaStream_t stream) { // @@ -492,7 +492,7 @@ void IVFPQ::search( int nprobe, int k, Tensor& outDistances, - Tensor& outIndices) { + Tensor& outIndices) { // These are caught at a higher level FAISS_ASSERT(nprobe <= GPU_MAX_SELECTION_K); FAISS_ASSERT(k <= GPU_MAX_SELECTION_K); @@ -509,7 +509,7 @@ void IVFPQ::search( resources_, makeTempAlloc(AllocType::Other, stream), {queries.getSize(0), nprobe}); - DeviceTensor coarseIndices( + DeviceTensor coarseIndices( resources_, makeTempAlloc(AllocType::Other, stream), {queries.getSize(0), nprobe}); @@ -537,10 +537,10 @@ void IVFPQ::searchPreassigned( Index* coarseQuantizer, Tensor& vecs, Tensor& ivfDistances, - Tensor& ivfAssignments, + Tensor& ivfAssignments, int k, Tensor& outDistances, - Tensor& outIndices, + Tensor& outIndices, bool storePairs) { FAISS_ASSERT(ivfDistances.getSize(0) == vecs.getSize(0)); FAISS_ASSERT(ivfAssignments.getSize(0) == vecs.getSize(0)); @@ -565,10 +565,10 @@ void IVFPQ::searchPreassigned( void IVFPQ::searchImpl_( Tensor& queries, Tensor& coarseDistances, - Tensor& coarseIndices, + Tensor& coarseIndices, int k, Tensor& outDistances, - Tensor& outIndices, + Tensor& outIndices, bool storePairs) { FAISS_ASSERT(storePairs == false); @@ -599,7 +599,7 @@ void IVFPQ::searchImpl_( // FIXME: we might ultimately be calling this function with inputs // from the CPU, these are unnecessary copies if (indicesOptions_ == INDICES_CPU) { - HostTensor hostOutIndices(outIndices, stream); + HostTensor hostOutIndices(outIndices, stream); ivfOffsetToUserIndex( hostOutIndices.data(), @@ -617,10 +617,10 @@ void IVFPQ::searchImpl_( void IVFPQ::runPQPrecomputedCodes_( Tensor& queries, Tensor& coarseDistances, - Tensor& coarseIndices, + Tensor& coarseIndices, int k, Tensor& outDistances, - Tensor& outIndices) { + Tensor& outIndices) { FAISS_ASSERT(metric_ == MetricType::METRIC_L2); auto stream = resources_->getDefaultStreamCurrentDevice(); @@ -705,10 +705,10 @@ void IVFPQ::runPQPrecomputedCodes_( void IVFPQ::runPQNoPrecomputedCodes_( Tensor& queries, Tensor& coarseDistances, - Tensor& coarseIndices, + Tensor& coarseIndices, int k, Tensor& outDistances, - Tensor& outIndices) { + Tensor& outIndices) { runPQScanMultiPassNoPrecomputed( queries, ivfCentroids_, diff --git a/faiss/gpu/impl/IVFPQ.cuh b/faiss/gpu/impl/IVFPQ.cuh index e13bccf8f..2990467c8 100644 --- a/faiss/gpu/impl/IVFPQ.cuh +++ b/faiss/gpu/impl/IVFPQ.cuh @@ -53,7 +53,7 @@ class IVFPQ : public IVFBase { int nprobe, int k, Tensor& outDistances, - Tensor& outIndices) override; + Tensor& outIndices) override; /// Performs search when we are already given the IVF cells to look at /// (GpuIndexIVF::search_preassigned implementation) @@ -61,10 +61,10 @@ class IVFPQ : public IVFBase { Index* coarseQuantizer, Tensor& vecs, Tensor& ivfDistances, - Tensor& ivfAssignments, + Tensor& ivfAssignments, int k, Tensor& outDistances, - Tensor& outIndices, + Tensor& outIndices, bool storePairs) override; protected: @@ -86,12 +86,12 @@ class IVFPQ : public IVFBase { void appendVectors_( Tensor& vecs, Tensor& ivfCentroidResiduals, - Tensor& indices, - Tensor& uniqueLists, + Tensor& indices, + Tensor& uniqueLists, Tensor& vectorsByUniqueList, Tensor& uniqueListVectorStart, Tensor& uniqueListStartOffset, - Tensor& listIds, + Tensor& listIds, Tensor& listOffset, cudaStream_t stream) override; @@ -100,10 +100,10 @@ class IVFPQ : public IVFBase { void searchImpl_( Tensor& queries, Tensor& coarseDistances, - Tensor& coarseIndices, + Tensor& coarseIndices, int k, Tensor& outDistances, - Tensor& outIndices, + Tensor& outIndices, bool storePairs); /// Sets the current product quantizer centroids; the data can be @@ -120,19 +120,19 @@ class IVFPQ : public IVFBase { void runPQPrecomputedCodes_( Tensor& queries, Tensor& coarseDistances, - Tensor& coarseIndices, + Tensor& coarseIndices, int k, Tensor& outDistances, - Tensor& outIndices); + Tensor& outIndices); /// Runs kernels for scanning inverted lists without precomputed codes void runPQNoPrecomputedCodes_( Tensor& queries, Tensor& coarseDistances, - Tensor& coarseIndices, + Tensor& coarseIndices, int k, Tensor& outDistances, - Tensor& outIndices); + Tensor& outIndices); private: /// Number of sub-quantizers per vector diff --git a/faiss/gpu/impl/IVFUtils.cu b/faiss/gpu/impl/IVFUtils.cu index 3bea02332..7e1e21814 100644 --- a/faiss/gpu/impl/IVFUtils.cu +++ b/faiss/gpu/impl/IVFUtils.cu @@ -21,7 +21,7 @@ namespace gpu { // Calculates the total number of intermediate distances to consider // for all queries __global__ void getResultLengths( - Tensor ivfListIds, + Tensor ivfListIds, int* listLengths, int totalSize, Tensor length) { @@ -34,7 +34,7 @@ __global__ void getResultLengths( int queryId = linearThreadId / nprobe; int listId = linearThreadId % nprobe; - Index::idx_t centroidId = ivfListIds[queryId][listId]; + idx_t centroidId = ivfListIds[queryId][listId]; // Safety guard in case NaNs in input cause no list ID to be generated length[queryId][listId] = (centroidId != -1) ? listLengths[centroidId] : 0; @@ -42,7 +42,7 @@ __global__ void getResultLengths( void runCalcListOffsets( GpuResources* res, - Tensor& ivfListIds, + Tensor& ivfListIds, DeviceVector& listLengths, Tensor& prefixSumOffsets, Tensor& thrustMem, diff --git a/faiss/gpu/impl/IVFUtils.cuh b/faiss/gpu/impl/IVFUtils.cuh index e7d7f7600..b5474af7e 100644 --- a/faiss/gpu/impl/IVFUtils.cuh +++ b/faiss/gpu/impl/IVFUtils.cuh @@ -23,7 +23,7 @@ class GpuResources; /// intermediate results for all (query, probe) pair void runCalcListOffsets( GpuResources* res, - Tensor& ivfListIds, + Tensor& ivfListIds, DeviceVector& listLengths, Tensor& prefixSumOffsets, Tensor& thrustMem, @@ -48,11 +48,11 @@ void runPass2SelectLists( DeviceVector& listIndices, IndicesOptions indicesOptions, Tensor& prefixSumOffsets, - Tensor& ivfListIds, + Tensor& ivfListIds, int k, bool chooseLargest, Tensor& outDistances, - Tensor& outIndices, + Tensor& outIndices, cudaStream_t stream); } // namespace gpu diff --git a/faiss/gpu/impl/IVFUtilsSelect2.cu b/faiss/gpu/impl/IVFUtilsSelect2.cu index 8977df59e..e008000bf 100644 --- a/faiss/gpu/impl/IVFUtilsSelect2.cu +++ b/faiss/gpu/impl/IVFUtilsSelect2.cu @@ -55,11 +55,11 @@ __global__ void pass2SelectLists( Tensor heapIndices, void** listIndices, Tensor prefixSumOffsets, - Tensor ivfListIds, + Tensor ivfListIds, int k, IndicesOptions opt, Tensor outDistances, - Tensor outIndices) { + Tensor outIndices) { constexpr int kNumWarps = ThreadsPerBlock / kWarpSize; __shared__ float smemK[kNumWarps * NumWarpQ]; @@ -110,7 +110,7 @@ __global__ void pass2SelectLists( // is the very last step and it is happening a small number of // times (#queries x k). int v = smemV[i]; - Index::idx_t index = -1; + idx_t index = -1; if (v != -1) { // `offset` is the offset of the intermediate result, as @@ -127,7 +127,7 @@ __global__ void pass2SelectLists( // This is then the probe for the query; we can find the actual // list ID from this - Index::idx_t listId = ivfListIds[queryId][probe]; + idx_t listId = ivfListIds[queryId][probe]; // Now, we need to know the offset within the list // We ensure that before the array (at offset -1), there is a 0 @@ -137,11 +137,11 @@ __global__ void pass2SelectLists( // This gives us our final index if (opt == INDICES_32_BIT) { - index = (Index::idx_t)((int*)listIndices[listId])[listOffset]; + index = (idx_t)((int*)listIndices[listId])[listOffset]; } else if (opt == INDICES_64_BIT) { - index = ((Index::idx_t*)listIndices[listId])[listOffset]; + index = ((idx_t*)listIndices[listId])[listOffset]; } else { - index = (listId << 32 | (Index::idx_t)listOffset); + index = (listId << 32 | (idx_t)listOffset); } } @@ -155,11 +155,11 @@ void runPass2SelectLists( DeviceVector& listIndices, IndicesOptions indicesOptions, Tensor& prefixSumOffsets, - Tensor& ivfListIds, + Tensor& ivfListIds, int k, bool chooseLargest, Tensor& outDistances, - Tensor& outIndices, + Tensor& outIndices, cudaStream_t stream) { auto grid = dim3(ivfListIds.getSize(0)); diff --git a/faiss/gpu/impl/IndexUtils.cu b/faiss/gpu/impl/IndexUtils.cu index 0f7499f7d..6e4f23448 100644 --- a/faiss/gpu/impl/IndexUtils.cu +++ b/faiss/gpu/impl/IndexUtils.cu @@ -22,25 +22,25 @@ int getMaxKSelection() { return GPU_MAX_SELECTION_K; } -void validateKSelect(Index::idx_t k) { +void validateKSelect(idx_t k) { FAISS_THROW_IF_NOT_FMT( - k > 0 && k < (Index::idx_t)getMaxKSelection(), + k > 0 && k < (idx_t)getMaxKSelection(), "GPU index only supports min/max-K selection up to %d (requested %zu)", getMaxKSelection(), k); } -void validateNProbe(Index::idx_t nprobe) { +void validateNProbe(idx_t nprobe) { FAISS_THROW_IF_NOT_FMT( - nprobe > 0 && nprobe < (Index::idx_t)getMaxKSelection(), + nprobe > 0 && nprobe < (idx_t)getMaxKSelection(), "GPU IVF index only supports nprobe selection up to %d (requested %zu)", getMaxKSelection(), nprobe); } -void validateNumVectors(Index::idx_t n) { +void validateNumVectors(idx_t n) { FAISS_THROW_IF_NOT_FMT( - n <= (Index::idx_t)std::numeric_limits::max(), + n <= (idx_t)std::numeric_limits::max(), "GPU index only supports up to %d indices (requested %zu)", std::numeric_limits::max(), n); diff --git a/faiss/gpu/impl/IndexUtils.h b/faiss/gpu/impl/IndexUtils.h index 9049792b0..d3d00c54c 100644 --- a/faiss/gpu/impl/IndexUtils.h +++ b/faiss/gpu/impl/IndexUtils.h @@ -20,13 +20,13 @@ namespace gpu { int getMaxKSelection(); // Validate the k parameter for search -void validateKSelect(Index::idx_t k); +void validateKSelect(idx_t k); // Validate the nprobe parameter for search -void validateNProbe(Index::idx_t nprobe); +void validateNProbe(idx_t nprobe); /// Validate the n (number of vectors) parameter for add, search, reconstruct -void validateNumVectors(Index::idx_t n); +void validateNumVectors(idx_t n); } // namespace gpu } // namespace faiss diff --git a/faiss/gpu/impl/PQCodeDistances-inl.cuh b/faiss/gpu/impl/PQCodeDistances-inl.cuh index a30afe886..2d5003449 100644 --- a/faiss/gpu/impl/PQCodeDistances-inl.cuh +++ b/faiss/gpu/impl/PQCodeDistances-inl.cuh @@ -32,7 +32,7 @@ __global__ void __launch_bounds__(288, 3) pqCodeDistances( int queriesPerBlock, Tensor coarseCentroids, Tensor pqCentroids, - Tensor coarseIndices, + Tensor coarseIndices, // (query id)(coarse)(subquantizer)(code) -> dist Tensor outCodeDistances) { const auto numSubQuantizers = pqCentroids.getSize(0); @@ -96,7 +96,7 @@ __global__ void __launch_bounds__(288, 3) pqCodeDistances( // Load list of coarse centroids found for (int i = threadIdx.x; i < coarseIndices.getSize(1); i += blockDim.x) { - // FIXME: coarseIndices is now Index::idx_t but the smem allocation + // FIXME: coarseIndices is now idx_t but the smem allocation // of coarseIds is still int. In practical limitation, everything // should still fit into int32 coarseIds[i] = (int)coarseIndices[queryId][i]; @@ -284,7 +284,7 @@ template __global__ void pqResidualVector( Tensor queries, Tensor coarseCentroids, - Tensor coarseIndices, + Tensor coarseIndices, int numSubDim, // output is transposed: // (sub q)(query id)(centroid id)(sub dim) @@ -292,7 +292,7 @@ __global__ void pqResidualVector( auto queryId = blockIdx.x; auto centroidId = blockIdx.y; - Index::idx_t realCentroidId = coarseIndices[queryId][centroidId]; + idx_t realCentroidId = coarseIndices[queryId][centroidId]; for (int dim = threadIdx.x; dim < queries.getSize(1); dim += blockDim.x) { float q = queries[queryId][dim]; @@ -323,7 +323,7 @@ void runPQResidualVector( Tensor& pqCentroids, Tensor& queries, Tensor& coarseCentroids, - Tensor& coarseIndices, + Tensor& coarseIndices, Tensor& residual, bool l2Residual, cudaStream_t stream) { @@ -401,7 +401,7 @@ void runPQCodeDistancesMM( Tensor& queries, Tensor& coarseCentroids, Tensor& coarseDistances, - Tensor& coarseIndices, + Tensor& coarseIndices, // Output is (query)(centroid)(sub q)(code) NoTypeTensor<4, true>& outCodeDistances, bool l2Distance, @@ -590,7 +590,7 @@ void runPQCodeDistances( Tensor& queries, Tensor& coarseCentroids, Tensor& coarseDistances, - Tensor& coarseIndices, + Tensor& coarseIndices, NoTypeTensor<4, true>& outCodeDistances, bool useMMImplementation, bool l2Distance, diff --git a/faiss/gpu/impl/PQCodeDistances.cuh b/faiss/gpu/impl/PQCodeDistances.cuh index 55c7d57ab..007e7c0d2 100644 --- a/faiss/gpu/impl/PQCodeDistances.cuh +++ b/faiss/gpu/impl/PQCodeDistances.cuh @@ -28,7 +28,7 @@ void runPQCodeDistances( Tensor& queries, Tensor& coarseCentroids, Tensor& coarseDistances, - Tensor& coarseIndices, + Tensor& coarseIndices, NoTypeTensor<4, true>& outCodeDistances, bool useMMImplementation, bool l2Distance, diff --git a/faiss/gpu/impl/PQScanMultiPassNoPrecomputed-inl.cuh b/faiss/gpu/impl/PQScanMultiPassNoPrecomputed-inl.cuh index 8c42438da..805782da7 100644 --- a/faiss/gpu/impl/PQScanMultiPassNoPrecomputed-inl.cuh +++ b/faiss/gpu/impl/PQScanMultiPassNoPrecomputed-inl.cuh @@ -28,7 +28,7 @@ template __global__ void pqScanInterleaved( Tensor queries, Tensor pqCentroids, - Tensor ivfListIds, + Tensor ivfListIds, Tensor codeDistances, void** listCodes, int* listLengths, @@ -38,7 +38,7 @@ __global__ void pqScanInterleaved( auto queryId = blockIdx.y; auto probeId = blockIdx.x; - Index::idx_t listId = ivfListIds[queryId][probeId]; + idx_t listId = ivfListIds[queryId][probeId]; // Safety guard in case NaNs in input cause no list ID to be generated if (listId == -1) { return; @@ -174,7 +174,7 @@ template __global__ void pqScanNoPrecomputedMultiPass( Tensor queries, Tensor pqCentroids, - Tensor ivfListIds, + Tensor ivfListIds, Tensor codeDistances, void** listCodes, int* listLengths, @@ -195,7 +195,7 @@ __global__ void pqScanNoPrecomputedMultiPass( int outBase = *(prefixSumOffsets[queryId][probeId].data() - 1); float* distanceOut = distance[outBase].data(); - Index::idx_t listId = ivfListIds[queryId][probeId]; + idx_t listId = ivfListIds[queryId][probeId]; // Safety guard in case NaNs in input cause no list ID to be generated if (listId == -1) { return; @@ -277,7 +277,7 @@ void runMultiPassTile( Tensor& pqCentroidsInnermostCode, NoTypeTensor<4, true>& codeDistances, Tensor& coarseDistances, - Tensor& coarseIndices, + Tensor& coarseIndices, bool useFloat16Lookup, bool useMMCodeDistance, bool interleavedCodeLayout, @@ -296,7 +296,7 @@ void runMultiPassTile( int k, faiss::MetricType metric, Tensor& outDistances, - Tensor& outIndices, + Tensor& outIndices, cudaStream_t stream) { // We only support two metrics at the moment FAISS_ASSERT( @@ -526,7 +526,7 @@ void runPQScanMultiPassNoPrecomputed( Tensor& centroids, Tensor& pqCentroidsInnermostCode, Tensor& coarseDistances, - Tensor& coarseIndices, + Tensor& coarseIndices, bool useFloat16Lookup, bool useMMCodeDistance, bool interleavedCodeLayout, @@ -543,7 +543,7 @@ void runPQScanMultiPassNoPrecomputed( // output Tensor& outDistances, // output - Tensor& outIndices, + Tensor& outIndices, GpuResources* res) { constexpr int kMinQueryTileSize = 8; constexpr int kMaxQueryTileSize = 65536; // typical max gridDim.y diff --git a/faiss/gpu/impl/PQScanMultiPassNoPrecomputed.cuh b/faiss/gpu/impl/PQScanMultiPassNoPrecomputed.cuh index f68959366..a8a6f02b9 100644 --- a/faiss/gpu/impl/PQScanMultiPassNoPrecomputed.cuh +++ b/faiss/gpu/impl/PQScanMultiPassNoPrecomputed.cuh @@ -24,7 +24,7 @@ void runPQScanMultiPassNoPrecomputed( Tensor& centroids, Tensor& pqCentroidsInnermostCode, Tensor& coarseDistances, - Tensor& coarseIndices, + Tensor& coarseIndices, bool useFloat16Lookup, bool useMMCodeDistance, bool interleavedCodeLayout, @@ -41,7 +41,7 @@ void runPQScanMultiPassNoPrecomputed( // output Tensor& outDistances, // output - Tensor& outIndices, + Tensor& outIndices, GpuResources* res); } // namespace gpu diff --git a/faiss/gpu/impl/PQScanMultiPassPrecomputed.cu b/faiss/gpu/impl/PQScanMultiPassPrecomputed.cu index aadf106a9..8c6e68be9 100644 --- a/faiss/gpu/impl/PQScanMultiPassPrecomputed.cu +++ b/faiss/gpu/impl/PQScanMultiPassPrecomputed.cu @@ -35,7 +35,7 @@ __global__ void pqScanPrecomputedInterleaved( Tensor precompTerm2, // (query id)(sub q)(code id) Tensor precompTerm3, - Tensor ivfListIds, + Tensor ivfListIds, void** listCodes, int* listLengths, Tensor prefixSumOffsets, @@ -44,7 +44,7 @@ __global__ void pqScanPrecomputedInterleaved( auto queryId = blockIdx.y; auto probeId = blockIdx.x; - Index::idx_t listId = ivfListIds[queryId][probeId]; + idx_t listId = ivfListIds[queryId][probeId]; // Safety guard in case NaNs in input cause no list ID to be generated if (listId == -1) { return; @@ -206,7 +206,7 @@ __global__ void pqScanPrecomputedMultiPass( Tensor precompTerm1, Tensor precompTerm2, Tensor precompTerm3, - Tensor ivfListIds, + Tensor ivfListIds, void** listCodes, int* listLengths, Tensor prefixSumOffsets, @@ -227,7 +227,7 @@ __global__ void pqScanPrecomputedMultiPass( int outBase = *(prefixSumOffsets[queryId][probeId].data() - 1); float* distanceOut = distance[outBase].data(); - Index::idx_t listId = ivfListIds[queryId][probeId]; + idx_t listId = ivfListIds[queryId][probeId]; // Safety guard in case NaNs in input cause no list ID to be generated if (listId == -1) { return; @@ -310,7 +310,7 @@ void runMultiPassTile( Tensor& precompTerm1, NoTypeTensor<3, true>& precompTerm2, NoTypeTensor<3, true>& precompTerm3, - Tensor& ivfListIds, + Tensor& ivfListIds, bool useFloat16Lookup, bool interleavedCodeLayout, int bitsPerSubQuantizer, @@ -327,7 +327,7 @@ void runMultiPassTile( Tensor& heapIndices, int k, Tensor& outDistances, - Tensor& outIndices, + Tensor& outIndices, cudaStream_t stream) { // Calculate offset lengths, so we know where to write out // intermediate results @@ -540,7 +540,7 @@ void runPQScanMultiPassPrecomputed( NoTypeTensor<3, true>& precompTerm2, // (query id)(sub q)(code id) NoTypeTensor<3, true>& precompTerm3, - Tensor& ivfListIds, + Tensor& ivfListIds, bool useFloat16Lookup, bool interleavedCodeLayout, int bitsPerSubQuantizer, @@ -555,7 +555,7 @@ void runPQScanMultiPassPrecomputed( // output Tensor& outDistances, // output - Tensor& outIndices, + Tensor& outIndices, GpuResources* res) { constexpr int kMinQueryTileSize = 8; constexpr int kMaxQueryTileSize = 65536; // typical max gridDim.y diff --git a/faiss/gpu/impl/PQScanMultiPassPrecomputed.cuh b/faiss/gpu/impl/PQScanMultiPassPrecomputed.cuh index bc49b771a..e7894d5ee 100644 --- a/faiss/gpu/impl/PQScanMultiPassPrecomputed.cuh +++ b/faiss/gpu/impl/PQScanMultiPassPrecomputed.cuh @@ -23,7 +23,7 @@ void runPQScanMultiPassPrecomputed( Tensor& precompTerm1, NoTypeTensor<3, true>& precompTerm2, NoTypeTensor<3, true>& precompTerm3, - Tensor& ivfListIds, + Tensor& ivfListIds, bool useFloat16Lookup, bool interleavedCodeLayout, int bitsPerSubQuantizer, @@ -38,7 +38,7 @@ void runPQScanMultiPassPrecomputed( // output Tensor& outDistances, // output - Tensor& outIndices, + Tensor& outIndices, GpuResources* res); } // namespace gpu diff --git a/faiss/gpu/impl/RemapIndices.cpp b/faiss/gpu/impl/RemapIndices.cpp index 321400c6e..8f78254e8 100644 --- a/faiss/gpu/impl/RemapIndices.cpp +++ b/faiss/gpu/impl/RemapIndices.cpp @@ -14,11 +14,11 @@ namespace gpu { // Utility function to translate (list id, offset) to a user index on // the CPU. In a cpp in order to use OpenMP void ivfOffsetToUserIndex( - Index::idx_t* indices, + idx_t* indices, int numLists, int queries, int k, - const std::vector>& listOffsetToUserIndex) { + const std::vector>& listOffsetToUserIndex) { FAISS_ASSERT(numLists == listOffsetToUserIndex.size()); #pragma omp parallel for diff --git a/faiss/gpu/impl/RemapIndices.h b/faiss/gpu/impl/RemapIndices.h index db8339594..34908b19e 100644 --- a/faiss/gpu/impl/RemapIndices.h +++ b/faiss/gpu/impl/RemapIndices.h @@ -16,11 +16,11 @@ namespace gpu { /// Utility function to translate (list id, offset) to a user index on /// the CPU. In a cpp in order to use OpenMP. void ivfOffsetToUserIndex( - Index::idx_t* indices, + idx_t* indices, int numLists, int queries, int k, - const std::vector>& listOffsetToUserIndex); + const std::vector>& listOffsetToUserIndex); } // namespace gpu } // namespace faiss diff --git a/faiss/gpu/impl/VectorResidual.cu b/faiss/gpu/impl/VectorResidual.cu index 8245bd67d..4d1563337 100644 --- a/faiss/gpu/impl/VectorResidual.cu +++ b/faiss/gpu/impl/VectorResidual.cu @@ -86,20 +86,20 @@ void calcResidual( void runCalcResidual( Tensor& vecs, Tensor& centroids, - Tensor& vecToCentroid, + Tensor& vecToCentroid, Tensor& residuals, cudaStream_t stream) { - calcResidual( + calcResidual( vecs, centroids, vecToCentroid, residuals, stream); } void runCalcResidual( Tensor& vecs, Tensor& centroids, - Tensor& vecToCentroid, + Tensor& vecToCentroid, Tensor& residuals, cudaStream_t stream) { - calcResidual( + calcResidual( vecs, centroids, vecToCentroid, residuals, stream); } @@ -190,38 +190,37 @@ void gatherReconstructByRange( } void runReconstruct( - Tensor& ids, + Tensor& ids, Tensor& vecs, Tensor& out, cudaStream_t stream) { - gatherReconstructByIds(ids, vecs, out, stream); + gatherReconstructByIds(ids, vecs, out, stream); } void runReconstruct( - Tensor& ids, + Tensor& ids, Tensor& vecs, Tensor& out, cudaStream_t stream) { - gatherReconstructByIds(ids, vecs, out, stream); + gatherReconstructByIds(ids, vecs, out, stream); } void runReconstruct( - Index::idx_t start, - Index::idx_t num, + idx_t start, + idx_t num, Tensor& vecs, Tensor& out, cudaStream_t stream) { - gatherReconstructByRange( - start, num, vecs, out, stream); + gatherReconstructByRange(start, num, vecs, out, stream); } void runReconstruct( - Index::idx_t start, - Index::idx_t num, + idx_t start, + idx_t num, Tensor& vecs, Tensor& out, cudaStream_t stream) { - gatherReconstructByRange(start, num, vecs, out, stream); + gatherReconstructByRange(start, num, vecs, out, stream); } } // namespace gpu diff --git a/faiss/gpu/impl/VectorResidual.cuh b/faiss/gpu/impl/VectorResidual.cuh index 9daf2366f..2803d188e 100644 --- a/faiss/gpu/impl/VectorResidual.cuh +++ b/faiss/gpu/impl/VectorResidual.cuh @@ -17,40 +17,40 @@ namespace gpu { void runCalcResidual( Tensor& vecs, Tensor& centroids, - Tensor& vecToCentroid, + Tensor& vecToCentroid, Tensor& residuals, cudaStream_t stream); void runCalcResidual( Tensor& vecs, Tensor& centroids, - Tensor& vecToCentroid, + Tensor& vecToCentroid, Tensor& residuals, cudaStream_t stream); // Gather vectors void runReconstruct( - Tensor& listIds, + Tensor& listIds, Tensor& vecs, Tensor& out, cudaStream_t stream); void runReconstruct( - Tensor& listIds, + Tensor& listIds, Tensor& vecs, Tensor& out, cudaStream_t stream); void runReconstruct( - Index::idx_t start, - Index::idx_t num, + idx_t start, + idx_t num, Tensor& vecs, Tensor& out, cudaStream_t stream); void runReconstruct( - Index::idx_t start, - Index::idx_t num, + idx_t start, + idx_t num, Tensor& vecs, Tensor& out, cudaStream_t stream); diff --git a/faiss/gpu/impl/scan/IVFInterleavedImpl.cuh b/faiss/gpu/impl/scan/IVFInterleavedImpl.cuh index 71073a7a0..7c873a3ce 100644 --- a/faiss/gpu/impl/scan/IVFInterleavedImpl.cuh +++ b/faiss/gpu/impl/scan/IVFInterleavedImpl.cuh @@ -15,7 +15,7 @@ \ void ivfInterleavedScanImpl_##WARP_Q##_( \ Tensor& queries, \ - Tensor& listIds, \ + Tensor& listIds, \ DeviceVector& listData, \ DeviceVector& listIndices, \ IndicesOptions indicesOptions, \ @@ -26,7 +26,7 @@ Tensor& residualBase, \ GpuScalarQuantizer* scalarQ, \ Tensor& outDistances, \ - Tensor& outIndices, \ + Tensor& outIndices, \ GpuResources* res) { \ FAISS_ASSERT(k <= WARP_Q); \ \ @@ -35,22 +35,22 @@ CUDA_TEST_ERROR(); \ } -#define IVF_INTERLEAVED_DECL(WARP_Q) \ - \ - void ivfInterleavedScanImpl_##WARP_Q##_( \ - Tensor& queries, \ - Tensor& listIds, \ - DeviceVector& listData, \ - DeviceVector& listIndices, \ - IndicesOptions indicesOptions, \ - DeviceVector& listLengths, \ - int k, \ - faiss::MetricType metric, \ - bool useResidual, \ - Tensor& residualBase, \ - GpuScalarQuantizer* scalarQ, \ - Tensor& outDistances, \ - Tensor& outIndices, \ +#define IVF_INTERLEAVED_DECL(WARP_Q) \ + \ + void ivfInterleavedScanImpl_##WARP_Q##_( \ + Tensor& queries, \ + Tensor& listIds, \ + DeviceVector& listData, \ + DeviceVector& listIndices, \ + IndicesOptions indicesOptions, \ + DeviceVector& listLengths, \ + int k, \ + faiss::MetricType metric, \ + bool useResidual, \ + Tensor& residualBase, \ + GpuScalarQuantizer* scalarQ, \ + Tensor& outDistances, \ + Tensor& outIndices, \ GpuResources* res) #define IVF_INTERLEAVED_CALL(WARP_Q) \ diff --git a/faiss/gpu/perf/PerfBinaryFlat.cu b/faiss/gpu/perf/PerfBinaryFlat.cu index fae7c1016..202cbce46 100644 --- a/faiss/gpu/perf/PerfBinaryFlat.cu +++ b/faiss/gpu/perf/PerfBinaryFlat.cu @@ -71,8 +71,7 @@ int main(int argc, char** argv) { // Time faiss CPU HostTensor cpuDistances({numQueries, FLAGS_k}); - HostTensor cpuIndices( - {numQueries, FLAGS_k}); + HostTensor cpuIndices({numQueries, FLAGS_k}); if (FLAGS_cpu) { float cpuTime = 0.0f; @@ -90,7 +89,7 @@ int main(int argc, char** argv) { } HostTensor gpuDistances({numQueries, FLAGS_k}); - HostTensor gpuIndices({numQueries, FLAGS_k}); + HostTensor gpuIndices({numQueries, FLAGS_k}); CUDA_VERIFY(cudaProfilerStart()); faiss::gpu::synchronizeAllDevices(); diff --git a/faiss/gpu/perf/PerfFlat.cu b/faiss/gpu/perf/PerfFlat.cu index bf71a5b9b..9db807a52 100644 --- a/faiss/gpu/perf/PerfFlat.cu +++ b/faiss/gpu/perf/PerfFlat.cu @@ -95,7 +95,7 @@ int main(int argc, char** argv) { // Time faiss CPU HostTensor cpuDistances({numQueries, FLAGS_k}); - HostTensor cpuIndices({numQueries, FLAGS_k}); + HostTensor cpuIndices({numQueries, FLAGS_k}); if (FLAGS_cpu) { float cpuTime = 0.0f; @@ -113,7 +113,7 @@ int main(int argc, char** argv) { } HostTensor gpuDistances({numQueries, FLAGS_k}); - HostTensor gpuIndices({numQueries, FLAGS_k}); + HostTensor gpuIndices({numQueries, FLAGS_k}); CUDA_VERIFY(cudaProfilerStart()); faiss::gpu::synchronizeAllDevices(); diff --git a/faiss/gpu/perf/PerfIVFFlat.cu b/faiss/gpu/perf/PerfIVFFlat.cu index 1b83d4f84..a3584ba38 100644 --- a/faiss/gpu/perf/PerfIVFFlat.cu +++ b/faiss/gpu/perf/PerfIVFFlat.cu @@ -93,7 +93,7 @@ int main(int argc, char** argv) { // Time faiss CPU HostTensor cpuDistances({numQueries, FLAGS_k}); - HostTensor cpuIndices({numQueries, FLAGS_k}); + HostTensor cpuIndices({numQueries, FLAGS_k}); float cpuTime = 0.0f; @@ -112,7 +112,7 @@ int main(int argc, char** argv) { printf("CPU time %.3f ms\n", cpuTime); HostTensor gpuDistances({numQueries, FLAGS_k}); - HostTensor gpuIndices({numQueries, FLAGS_k}); + HostTensor gpuIndices({numQueries, FLAGS_k}); CUDA_VERIFY(cudaProfilerStart()); faiss::gpu::synchronizeAllDevices(); diff --git a/faiss/gpu/perf/PerfIVFPQ.cu b/faiss/gpu/perf/PerfIVFPQ.cu index 3665c1e10..baae8af23 100644 --- a/faiss/gpu/perf/PerfIVFPQ.cu +++ b/faiss/gpu/perf/PerfIVFPQ.cu @@ -103,7 +103,7 @@ int main(int argc, char** argv) { // Time faiss CPU HostTensor cpuDistances({numQueries, FLAGS_k}); - HostTensor cpuIndices({numQueries, FLAGS_k}); + HostTensor cpuIndices({numQueries, FLAGS_k}); float cpuTime = 0.0f; @@ -122,7 +122,7 @@ int main(int argc, char** argv) { printf("CPU time %.3f ms\n", cpuTime); HostTensor gpuDistances({numQueries, FLAGS_k}); - HostTensor gpuIndices({numQueries, FLAGS_k}); + HostTensor gpuIndices({numQueries, FLAGS_k}); CUDA_VERIFY(cudaProfilerStart()); faiss::gpu::synchronizeAllDevices(); diff --git a/faiss/gpu/test/TestGpuDistance.cu b/faiss/gpu/test/TestGpuDistance.cu index 2672496ee..cb9b1a417 100644 --- a/faiss/gpu/test/TestGpuDistance.cu +++ b/faiss/gpu/test/TestGpuDistance.cu @@ -60,7 +60,7 @@ void testTransposition( cpuIndex.add(numVecs, vecs.data()); std::vector cpuDistance(numQuery * k, 0); - std::vector cpuIndices(numQuery * k, -1); + std::vector cpuIndices(numQuery * k, -1); cpuIndex.search( numQuery, queries.data(), k, cpuDistance.data(), cpuIndices.data()); @@ -97,7 +97,7 @@ void testTransposition( runTransposeAny(gpuQueries, 0, 1, queriesT, stream); std::vector gpuDistance(numQuery * k, 0); - std::vector gpuIndices(numQuery * k, -1); + std::vector gpuIndices(numQuery * k, -1); GpuDistanceParams args; args.metric = metric; diff --git a/faiss/gpu/test/TestGpuIndexBinaryFlat.cpp b/faiss/gpu/test/TestGpuIndexBinaryFlat.cpp index 0876e12a8..8b8e8f264 100644 --- a/faiss/gpu/test/TestGpuIndexBinaryFlat.cpp +++ b/faiss/gpu/test/TestGpuIndexBinaryFlat.cpp @@ -18,9 +18,9 @@ void compareBinaryDist( const std::vector& cpuDist, - const std::vector& cpuLabels, + const std::vector& cpuLabels, const std::vector& gpuDist, - const std::vector& gpuLabels, + const std::vector& gpuLabels, int numQuery, int k) { for (int i = 0; i < numQuery; ++i) { @@ -29,8 +29,8 @@ void compareBinaryDist( // encounters the values. The last set of equivalent distances seen in // the min-k might be truncated, so we can't check that set, but all // others we can check. - std::set cpuLabelSet; - std::set gpuLabelSet; + std::set cpuLabelSet; + std::set gpuLabelSet; int curDist = -1; @@ -89,13 +89,13 @@ void testGpuIndexBinaryFlat(int kOverride = -1) { auto query = faiss::gpu::randBinaryVecs(numQuery, dims); std::vector cpuDist(numQuery * k); - std::vector cpuLabels(numQuery * k); + std::vector cpuLabels(numQuery * k); cpuIndex.search( numQuery, query.data(), k, cpuDist.data(), cpuLabels.data()); std::vector gpuDist(numQuery * k); - std::vector gpuLabels(numQuery * k); + std::vector gpuLabels(numQuery * k); gpuIndex.search( numQuery, query.data(), k, gpuDist.data(), gpuLabels.data()); diff --git a/faiss/gpu/test/TestGpuIndexFlat.cpp b/faiss/gpu/test/TestGpuIndexFlat.cpp index b727a15ef..617db62d9 100644 --- a/faiss/gpu/test/TestGpuIndexFlat.cpp +++ b/faiss/gpu/test/TestGpuIndexFlat.cpp @@ -220,7 +220,7 @@ TEST(TestGpuIndexFlat, QueryEmpty) { std::vector queries(numQuery * dim, 1.0f); std::vector dist(numQuery * k, 0); - std::vector ind(numQuery * k); + std::vector ind(numQuery * k); gpuIndex.search(numQuery, queries.data(), k, dist.data(), ind.data()); @@ -437,7 +437,7 @@ TEST(TestGpuIndexFlat, Residual) { cpuIndex.add(numVecs, vecs.data()); gpuIndex.add(numVecs, vecs.data()); - auto indexVecs = std::vector{0, 2, 4, 6, 8}; + auto indexVecs = std::vector{0, 2, 4, 6, 8}; auto queryVecs = faiss::gpu::randVecs(indexVecs.size(), dim); auto residualsCpu = std::vector(indexVecs.size() * dim); @@ -517,7 +517,7 @@ TEST(TestGpuIndexFlat, Reconstruct) { // Test reconstruct_batch if (false) { - auto reconstructKeys = std::vector{1, 3, 5}; + auto reconstructKeys = std::vector{1, 3, 5}; auto reconstructVecs = std::vector(reconstructKeys.size() * dim); @@ -565,7 +565,7 @@ TEST(TestGpuIndexFlat, SearchAndReconstruct) { gpuIndex.add(nb, xb.data()); std::vector refDistance(nq * k, 0); - std::vector refIndices(nq * k, -1); + std::vector refIndices(nq * k, -1); std::vector refReconstruct(nq * k * dim, 0); cpuIndex.search_and_reconstruct( nq, @@ -576,7 +576,7 @@ TEST(TestGpuIndexFlat, SearchAndReconstruct) { refReconstruct.data()); std::vector testDistance(nq * k, 0); - std::vector testIndices(nq * k, -1); + std::vector testIndices(nq * k, -1); std::vector testReconstruct(nq * k * dim, 0); gpuIndex.search_and_reconstruct( nq, @@ -606,7 +606,7 @@ TEST(TestGpuIndexFlat, SearchAndReconstruct) { // above will ensure a decent number of matches), reconstruction should be // the same for the vectors that do match for (int i = 0; i < nq; ++i) { - std::unordered_map refLocation; + std::unordered_map refLocation; for (int j = 0; j < k; ++j) { refLocation.insert(std::make_pair(refIndices[i * k + j], j)); diff --git a/faiss/gpu/test/TestGpuIndexIVFFlat.cpp b/faiss/gpu/test/TestGpuIndexIVFFlat.cpp index 086959379..ca61fc3f1 100644 --- a/faiss/gpu/test/TestGpuIndexIVFFlat.cpp +++ b/faiss/gpu/test/TestGpuIndexIVFFlat.cpp @@ -441,7 +441,7 @@ TEST(TestGpuIndexIVFFlat, QueryNaN) { numQuery * opt.dim, std::numeric_limits::quiet_NaN()); std::vector distances(numQuery * opt.k, 0); - std::vector indices(numQuery * opt.k, 0); + std::vector indices(numQuery * opt.k, 0); gpuIndex.search( numQuery, nans.data(), opt.k, distances.data(), indices.data()); @@ -490,7 +490,7 @@ TEST(TestGpuIndexIVFFlat, AddNaN) { std::vector queryVecs = faiss::gpu::randVecs(opt.numQuery, opt.dim); std::vector distance(opt.numQuery * opt.k, 0); - std::vector indices(opt.numQuery * opt.k, 0); + std::vector indices(opt.numQuery * opt.k, 0); // should not crash gpuIndex.search( diff --git a/faiss/gpu/test/TestGpuIndexIVFPQ.cpp b/faiss/gpu/test/TestGpuIndexIVFPQ.cpp index 64459ca6c..c7c23a8e4 100644 --- a/faiss/gpu/test/TestGpuIndexIVFPQ.cpp +++ b/faiss/gpu/test/TestGpuIndexIVFPQ.cpp @@ -655,7 +655,7 @@ TEST(TestGpuIndexIVFPQ, QueryNaN) { numQuery * opt.dim, std::numeric_limits::quiet_NaN()); std::vector distances(numQuery * opt.k, 0); - std::vector indices(numQuery * opt.k, 0); + std::vector indices(numQuery * opt.k, 0); gpuIndex.search( numQuery, nans.data(), opt.k, distances.data(), indices.data()); @@ -711,7 +711,7 @@ TEST(TestGpuIndexIVFPQ, AddNaN) { std::vector queryVecs = faiss::gpu::randVecs(opt.numQuery, opt.dim); std::vector distance(opt.numQuery * opt.k, 0); - std::vector indices(opt.numQuery * opt.k, 0); + std::vector indices(opt.numQuery * opt.k, 0); // should not crash gpuIndex.search( diff --git a/faiss/gpu/test/TestUtils.cpp b/faiss/gpu/test/TestUtils.cpp index 552040309..c81d34339 100644 --- a/faiss/gpu/test/TestUtils.cpp +++ b/faiss/gpu/test/TestUtils.cpp @@ -97,7 +97,7 @@ void compareIndices( float pctMaxDiffN) { // Compare std::vector refDistance(numQuery * k, 0); - std::vector refIndices(numQuery * k, -1); + std::vector refIndices(numQuery * k, -1); refIndex.search( numQuery, queryVecs.data(), @@ -106,7 +106,7 @@ void compareIndices( refIndices.data()); std::vector testDistance(numQuery * k, 0); - std::vector testIndices(numQuery * k, -1); + std::vector testIndices(numQuery * k, -1); testIndex.search( numQuery, queryVecs.data(), @@ -162,9 +162,9 @@ inline T lookup(const T* p, int i, int j, int /*dim1*/, int dim2) { void compareLists( const float* refDist, - const faiss::Index::idx_t* refInd, + const faiss::idx_t* refInd, const float* testDist, - const faiss::Index::idx_t* testInd, + const faiss::idx_t* testInd, int dim1, int dim2, const std::string& configMsg, @@ -181,10 +181,10 @@ void compareLists( int numResults = dim1 * dim2; // query -> {index -> result position} - std::vector> refIndexMap; + std::vector> refIndexMap; for (int query = 0; query < dim1; ++query) { - std::unordered_map indices; + std::unordered_map indices; for (int result = 0; result < dim2; ++result) { indices[lookup(refInd, query, result, dim1, dim2)] = result; @@ -208,7 +208,7 @@ void compareLists( for (int query = 0; query < dim1; ++query) { std::vector diffs; - std::set uniqueIndices; + std::set uniqueIndices; auto& indices = refIndexMap[query]; diff --git a/faiss/gpu/test/TestUtils.h b/faiss/gpu/test/TestUtils.h index 2491b4f58..3edd1daef 100644 --- a/faiss/gpu/test/TestUtils.h +++ b/faiss/gpu/test/TestUtils.h @@ -93,9 +93,9 @@ void compareIndices( /// Display specific differences in the two (distance, index) lists void compareLists( const float* refDist, - const faiss::Index::idx_t* refInd, + const faiss::idx_t* refInd, const float* testDist, - const faiss::Index::idx_t* testInd, + const faiss::idx_t* testInd, int dim1, int dim2, const std::string& configMsg, @@ -130,13 +130,13 @@ void testIVFEquality(A& cpuIndex, B& gpuIndex) { EXPECT_EQ(cpuCodes, gpuCodes); // Index equality - std::vector cpuIndices(cpuLists->list_size(i)); + std::vector cpuIndices(cpuLists->list_size(i)); auto si = faiss::InvertedLists::ScopedIds(cpuLists, i); std::memcpy( cpuIndices.data(), si.get(), - cpuLists->list_size(i) * sizeof(faiss::Index::idx_t)); + cpuLists->list_size(i) * sizeof(faiss::idx_t)); EXPECT_EQ(cpuIndices, gpuIndex.getListIndices(i)); } } diff --git a/faiss/gpu/test/demo_ivfpq_indexing_gpu.cpp b/faiss/gpu/test/demo_ivfpq_indexing_gpu.cpp index 10a1f760a..1dd52bf0d 100644 --- a/faiss/gpu/test/demo_ivfpq_indexing_gpu.cpp +++ b/faiss/gpu/test/demo_ivfpq_indexing_gpu.cpp @@ -130,7 +130,7 @@ int main() { k, nq); - std::vector nns(k * nq); + std::vector nns(k * nq); std::vector dis(k * nq); index.search(nq, queries.data(), k, dis.data(), nns.data()); diff --git a/faiss/impl/AdditiveQuantizer.h b/faiss/impl/AdditiveQuantizer.h index 4b4c8717c..518a2c7b4 100644 --- a/faiss/impl/AdditiveQuantizer.h +++ b/faiss/impl/AdditiveQuantizer.h @@ -157,7 +157,6 @@ struct AdditiveQuantizer : Quantizer { * Support for exhaustive distance computations with all the centroids. * Hence, the number of these centroids should not be too large. ****************************************************************************/ - using idx_t = Index::idx_t; /// decoding function for a code in a 64-bit word void decode_64bit(idx_t n, float* x) const; diff --git a/faiss/impl/AuxIndexStructures.h b/faiss/impl/AuxIndexStructures.h index b88903f19..6c878d57e 100644 --- a/faiss/impl/AuxIndexStructures.h +++ b/faiss/impl/AuxIndexStructures.h @@ -18,7 +18,7 @@ #include #include -#include +#include #include namespace faiss { @@ -31,8 +31,6 @@ struct RangeSearchResult { size_t nq; ///< nb of queries size_t* lims; ///< size (nq + 1) - typedef Index::idx_t idx_t; - idx_t* labels; ///< result for query i is labels[lims[i]:lims[i+1]] float* distances; ///< corresponding distances (not sorted) @@ -62,8 +60,6 @@ struct RangeSearchResult { /** List of temporary buffers used to store results before they are * copied to the RangeSearchResult object. */ struct BufferList { - typedef Index::idx_t idx_t; - // buffer sizes in # entries size_t buffer_size; @@ -94,7 +90,6 @@ struct RangeSearchPartialResult; /// result structure for a single query struct RangeQueryResult { - using idx_t = Index::idx_t; idx_t qno; //< id of the query size_t nres; //< nb of results for this query RangeSearchPartialResult* pres; diff --git a/faiss/impl/DistanceComputer.h b/faiss/impl/DistanceComputer.h index 5942790ed..9e7f248d0 100644 --- a/faiss/impl/DistanceComputer.h +++ b/faiss/impl/DistanceComputer.h @@ -23,8 +23,6 @@ namespace faiss { * that has additional methods to handle the inverted list context. ***********************************************************/ struct DistanceComputer { - using idx_t = Index::idx_t; - /// called before computing distances. Pointer x should remain valid /// while operator () is called virtual void set_query(const float* x) = 0; diff --git a/faiss/impl/HNSW.cpp b/faiss/impl/HNSW.cpp index 1ed15d790..12ae55a7c 100644 --- a/faiss/impl/HNSW.cpp +++ b/faiss/impl/HNSW.cpp @@ -509,7 +509,6 @@ void HNSW::add_with_locks( namespace { -using idx_t = HNSW::idx_t; using MinimaxHeap = HNSW::MinimaxHeap; using Node = HNSW::Node; /** Do a BFS on the candidates list */ diff --git a/faiss/impl/HNSW.h b/faiss/impl/HNSW.h index 21cc94eeb..c3cff678a 100644 --- a/faiss/impl/HNSW.h +++ b/faiss/impl/HNSW.h @@ -52,10 +52,7 @@ struct SearchParametersHNSW : SearchParameters { struct HNSW { /// internal storage of vectors (32 bits: this is expensive) - typedef int storage_idx_t; - - /// Faiss results are 64-bit - typedef Index::idx_t idx_t; + using storage_idx_t = int32_t; typedef std::pair Node; diff --git a/faiss/impl/IDSelector.cpp b/faiss/impl/IDSelector.cpp index 8810d3850..e4a4bba96 100644 --- a/faiss/impl/IDSelector.cpp +++ b/faiss/impl/IDSelector.cpp @@ -92,7 +92,7 @@ IDSelectorBatch::IDSelectorBatch(size_t n, const idx_t* indices) { mask = ((idx_t)1 << nbits) - 1; bloom.resize((idx_t)1 << (nbits - 3), 0); for (idx_t i = 0; i < n; i++) { - Index::idx_t id = indices[i]; + idx_t id = indices[i]; set.insert(id); id &= mask; bloom[id >> 3] |= 1 << (id & 7); diff --git a/faiss/impl/IDSelector.h b/faiss/impl/IDSelector.h index 449448971..e00dbe8fb 100644 --- a/faiss/impl/IDSelector.h +++ b/faiss/impl/IDSelector.h @@ -19,7 +19,6 @@ namespace faiss { /** Encapsulates a set of ids to handle. */ struct IDSelector { - using idx_t = Index::idx_t; virtual bool is_member(idx_t id) const = 0; virtual ~IDSelector() {} }; diff --git a/faiss/impl/NNDescent.h b/faiss/impl/NNDescent.h index 858367f3b..74982dbc0 100644 --- a/faiss/impl/NNDescent.h +++ b/faiss/impl/NNDescent.h @@ -90,7 +90,6 @@ struct Nhood { struct NNDescent { using storage_idx_t = int; - using idx_t = Index::idx_t; using KNNGraph = std::vector; diff --git a/faiss/impl/NSG.cpp b/faiss/impl/NSG.cpp index f19caf01e..6dbf0c921 100644 --- a/faiss/impl/NSG.cpp +++ b/faiss/impl/NSG.cpp @@ -29,8 +29,6 @@ constexpr int EMPTY_ID = -1; distances. This makes supporting INNER_PRODUCE search easier */ struct NegativeDistanceComputer : DistanceComputer { - using idx_t = Index::idx_t; - /// owned by this DistanceComputer* basedis; diff --git a/faiss/impl/NSG.h b/faiss/impl/NSG.h index 132c0050f..c26c9b291 100644 --- a/faiss/impl/NSG.h +++ b/faiss/impl/NSG.h @@ -98,10 +98,7 @@ DistanceComputer* storage_distance_computer(const Index* storage); struct NSG { /// internal storage of vectors (32 bits: this is expensive) - using storage_idx_t = int; - - /// Faiss results are 64-bit - using idx_t = Index::idx_t; + using storage_idx_t = int32_t; int ntotal; ///< nb of nodes diff --git a/faiss/impl/ProductQuantizer.h b/faiss/impl/ProductQuantizer.h index e5c1336b4..5980cfdbd 100644 --- a/faiss/impl/ProductQuantizer.h +++ b/faiss/impl/ProductQuantizer.h @@ -23,8 +23,6 @@ namespace faiss { /** Product Quantizer. Implemented only for METRIC_L2 */ struct ProductQuantizer : Quantizer { - using idx_t = Index::idx_t; - size_t M; ///< number of subquantizers size_t nbits; ///< number of bits per quantization index diff --git a/faiss/impl/Quantizer.h b/faiss/impl/Quantizer.h index f0750a134..9ee86c00a 100644 --- a/faiss/impl/Quantizer.h +++ b/faiss/impl/Quantizer.h @@ -8,8 +8,6 @@ namespace faiss { /** Product Quantizer. Implemented only for METRIC_L2 */ struct Quantizer { - using idx_t = Index::idx_t; - size_t d; ///< size of the input vectors size_t code_size; ///< bytes per indexed vector diff --git a/faiss/impl/ResidualQuantizer.cpp b/faiss/impl/ResidualQuantizer.cpp index 3596ba157..6eb747805 100644 --- a/faiss/impl/ResidualQuantizer.cpp +++ b/faiss/impl/ResidualQuantizer.cpp @@ -143,8 +143,6 @@ void beam_search_encode_step( // we have to fill in the whole output matrix FAISS_THROW_IF_NOT(new_beam_size <= beam_size * K); - using idx_t = Index::idx_t; - std::vector cent_distances; std::vector cent_ids; @@ -544,8 +542,7 @@ size_t ResidualQuantizer::memory_per_point(int beam_size) const { size_t mem; mem = beam_size * d * 2 * sizeof(float); // size for 2 beams at a time mem += beam_size * beam_size * - (sizeof(float) + - sizeof(Index::idx_t)); // size for 1 beam search result + (sizeof(float) + sizeof(idx_t)); // size for 1 beam search result return mem; } diff --git a/faiss/impl/ScalarQuantizer.cpp b/faiss/impl/ScalarQuantizer.cpp index 80546767e..34e2f4e6d 100644 --- a/faiss/impl/ScalarQuantizer.cpp +++ b/faiss/impl/ScalarQuantizer.cpp @@ -54,7 +54,6 @@ namespace faiss { namespace { -typedef Index::idx_t idx_t; typedef ScalarQuantizer::QuantizerType QuantizerType; typedef ScalarQuantizer::RangeStat RangeStat; using SQDistanceComputer = ScalarQuantizer::SQDistanceComputer; @@ -1131,7 +1130,7 @@ void ScalarQuantizer::train_residual( ScopeDeleter del_x(x_in == x ? nullptr : x); if (by_residual) { - std::vector idx(n); + std::vector idx(n); quantizer->assign(n, x, idx.data()); std::vector residuals(n * d); diff --git a/faiss/impl/index_read.cpp b/faiss/impl/index_read.cpp index 740ac94a3..f38c7980d 100644 --- a/faiss/impl/index_read.cpp +++ b/faiss/impl/index_read.cpp @@ -65,7 +65,7 @@ namespace faiss { static void read_index_header(Index* idx, IOReader* f) { READ1(idx->d); READ1(idx->ntotal); - Index::idx_t dummy; + idx_t dummy; READ1(dummy); READ1(dummy); READ1(idx->is_trained); @@ -439,7 +439,6 @@ static void read_direct_map(DirectMap* dm, IOReader* f) { dm->type = (DirectMap::Type)maintain_direct_map; READVECTOR(dm->array); if (dm->type == DirectMap::Hashtable) { - using idx_t = Index::idx_t; std::vector> v; READVECTOR(v); std::unordered_map& map = dm->hashtable; @@ -453,7 +452,7 @@ static void read_direct_map(DirectMap* dm, IOReader* f) { static void read_ivf_header( IndexIVF* ivf, IOReader* f, - std::vector>* ids = nullptr) { + std::vector>* ids = nullptr) { read_index_header(ivf, f); READ1(ivf->nlist); READ1(ivf->nprobe); @@ -470,7 +469,7 @@ static void read_ivf_header( // used for legacy formats static ArrayInvertedLists* set_array_invlist( IndexIVF* ivf, - std::vector>& ids) { + std::vector>& ids) { ArrayInvertedLists* ail = new ArrayInvertedLists(ivf->nlist, ivf->code_size); std::swap(ail->ids, ids); @@ -487,7 +486,7 @@ static IndexIVFPQ* read_ivfpq(IOReader* f, uint32_t h, int io_flags) { : nullptr; IndexIVFPQ* ivpq = ivfpqr ? ivfpqr : new IndexIVFPQ(); - std::vector> ids; + std::vector> ids; read_ivf_header(ivpq, f, legacy ? &ids : nullptr); READ1(ivpq->by_residual); READ1(ivpq->code_size); @@ -731,7 +730,7 @@ Index* read_index(IOReader* f, int io_flags) { idx = ivaqfs; } else if (h == fourcc("IvFl") || h == fourcc("IvFL")) { // legacy IndexIVFFlat* ivfl = new IndexIVFFlat(); - std::vector> ids; + std::vector> ids; read_ivf_header(ivfl, f, &ids); ivfl->code_size = ivfl->d * sizeof(float); ArrayInvertedLists* ail = set_array_invlist(ivfl, ids); @@ -754,10 +753,10 @@ Index* read_index(IOReader* f, int io_flags) { read_ivf_header(ivfl, f); ivfl->code_size = ivfl->d * sizeof(float); { - std::vector tab; + std::vector tab; READVECTOR(tab); for (long i = 0; i < tab.size(); i += 2) { - std::pair pair(tab[i], tab[i + 1]); + std::pair pair(tab[i], tab[i + 1]); ivfl->instances.insert(pair); } } @@ -788,7 +787,7 @@ Index* read_index(IOReader* f, int io_flags) { idx = idxl; } else if (h == fourcc("IvSQ")) { // legacy IndexIVFScalarQuantizer* ivsc = new IndexIVFScalarQuantizer(); - std::vector> ids; + std::vector> ids; read_ivf_header(ivsc, f, &ids); read_ScalarQuantizer(&ivsc->sq, f); READ1(ivsc->code_size); @@ -1072,7 +1071,7 @@ static void read_index_binary_header(IndexBinary* idx, IOReader* f) { static void read_binary_ivf_header( IndexBinaryIVF* ivf, IOReader* f, - std::vector>* ids = nullptr) { + std::vector>* ids = nullptr) { read_index_binary_header(ivf, f); READ1(ivf->nlist); READ1(ivf->nprobe); diff --git a/faiss/impl/index_write.cpp b/faiss/impl/index_write.cpp index 84e9f964c..6ea40e1be 100644 --- a/faiss/impl/index_write.cpp +++ b/faiss/impl/index_write.cpp @@ -84,7 +84,7 @@ namespace faiss { static void write_index_header(const Index* idx, IOWriter* f) { WRITE1(idx->d); WRITE1(idx->ntotal); - Index::idx_t dummy = 1 << 20; + idx_t dummy = 1 << 20; WRITE1(dummy); WRITE1(dummy); WRITE1(idx->is_trained); @@ -373,7 +373,6 @@ static void write_direct_map(const DirectMap* dm, IOWriter* f) { WRITE1(maintain_direct_map); WRITEVECTOR(dm->array); if (dm->type == DirectMap::Hashtable) { - using idx_t = Index::idx_t; std::vector> v; const std::unordered_map& map = dm->hashtable; v.resize(map.size()); @@ -615,7 +614,7 @@ void write_index(const Index* idx, IOWriter* f) { WRITE1(h); write_ivf_header(ivfl, f); { - std::vector tab(2 * ivfl->instances.size()); + std::vector tab(2 * ivfl->instances.size()); long i = 0; for (auto it = ivfl->instances.begin(); it != ivfl->instances.end(); ++it) { @@ -900,7 +899,7 @@ static void write_binary_multi_hash_map( size_t ntotal, IOWriter* f) { int id_bits = 0; - while ((ntotal > ((Index::idx_t)1 << id_bits))) { + while ((ntotal > ((idx_t)1 << id_bits))) { id_bits++; } WRITE1(id_bits); diff --git a/faiss/impl/kmeans1d.cpp b/faiss/impl/kmeans1d.cpp index 82e5c5d68..faa4fad84 100644 --- a/faiss/impl/kmeans1d.cpp +++ b/faiss/impl/kmeans1d.cpp @@ -20,7 +20,6 @@ namespace faiss { -using idx_t = Index::idx_t; using LookUpFunc = std::function; void reduce( diff --git a/faiss/impl/kmeans1d.h b/faiss/impl/kmeans1d.h index 0e8a1994f..f9a132668 100644 --- a/faiss/impl/kmeans1d.h +++ b/faiss/impl/kmeans1d.h @@ -22,10 +22,10 @@ namespace faiss { * @param argmins argmin of each row */ void smawk( - const Index::idx_t nrows, - const Index::idx_t ncols, + const idx_t nrows, + const idx_t ncols, const float* x, - Index::idx_t* argmins); + idx_t* argmins); /** Exact 1D K-Means by dynamic programming * diff --git a/faiss/impl/platform_macros.h b/faiss/impl/platform_macros.h index 8bf4b9ff6..c792ac39a 100644 --- a/faiss/impl/platform_macros.h +++ b/faiss/impl/platform_macros.h @@ -7,6 +7,10 @@ #pragma once +// basic int types and size_t +#include +#include + #ifdef _MSC_VER /******************************************************* @@ -87,3 +91,9 @@ inline int __builtin_clzll(uint64_t x) { #define ALIGNED(x) __attribute__((aligned(x))) #endif // _MSC_VER + +#if defined(__GNUC__) || defined(__clang__) +#define FAISS_DEPRECATED(msg) __attribute__((deprecated(msg))) +#else +#define FAISS_DEPRECATED(msg) +#endif // GCC or Clang diff --git a/faiss/invlists/BlockInvertedLists.cpp b/faiss/invlists/BlockInvertedLists.cpp index 58345ff24..095231dd4 100644 --- a/faiss/invlists/BlockInvertedLists.cpp +++ b/faiss/invlists/BlockInvertedLists.cpp @@ -61,7 +61,7 @@ const uint8_t* BlockInvertedLists::get_codes(size_t list_no) const { return codes[list_no].get(); } -const InvertedLists::idx_t* BlockInvertedLists::get_ids(size_t list_no) const { +const idx_t* BlockInvertedLists::get_ids(size_t list_no) const { assert(list_no < nlist); return ids[list_no].data(); } diff --git a/faiss/invlists/DirectMap.cpp b/faiss/invlists/DirectMap.cpp index 6cb520225..b276b764e 100644 --- a/faiss/invlists/DirectMap.cpp +++ b/faiss/invlists/DirectMap.cpp @@ -68,7 +68,7 @@ void DirectMap::clear() { hashtable.clear(); } -DirectMap::idx_t DirectMap::get(idx_t key) const { +idx_t DirectMap::get(idx_t key) const { if (type == Array) { FAISS_THROW_IF_NOT_MSG(key >= 0 && key < array.size(), "invalid key"); idx_t lo = array[key]; diff --git a/faiss/invlists/DirectMap.h b/faiss/invlists/DirectMap.h index 15e8b8f70..30117d6d4 100644 --- a/faiss/invlists/DirectMap.h +++ b/faiss/invlists/DirectMap.h @@ -15,6 +15,8 @@ namespace faiss { +struct IDSelector; + // When offsets list id + offset are encoded in an uint64 // we call this LO = list-offset @@ -34,8 +36,6 @@ inline uint64_t lo_offset(uint64_t lo) { * Direct map: a way to map back from ids to inverted lists */ struct DirectMap { - typedef Index::idx_t idx_t; - enum Type { NoMap = 0, // default Array = 1, // sequential ids (only for add, no add_with_ids) @@ -91,8 +91,6 @@ struct DirectMap { /// Thread-safe way of updating the direct_map struct DirectMapAdd { - typedef Index::idx_t idx_t; - using Type = DirectMap::Type; DirectMap& direct_map; diff --git a/faiss/invlists/InvertedLists.cpp b/faiss/invlists/InvertedLists.cpp index 2defe8143..219f42ac4 100644 --- a/faiss/invlists/InvertedLists.cpp +++ b/faiss/invlists/InvertedLists.cpp @@ -25,8 +25,7 @@ InvertedLists::InvertedLists(size_t nlist, size_t code_size) InvertedLists::~InvertedLists() {} -InvertedLists::idx_t InvertedLists::get_single_id(size_t list_no, size_t offset) - const { +idx_t InvertedLists::get_single_id(size_t list_no, size_t offset) const { assert(offset < list_size(list_no)); const idx_t* ids = get_ids(list_no); idx_t id = ids[offset]; @@ -158,7 +157,7 @@ const uint8_t* ArrayInvertedLists::get_codes(size_t list_no) const { return codes[list_no].data(); } -const InvertedLists::idx_t* ArrayInvertedLists::get_ids(size_t list_no) const { +const idx_t* ArrayInvertedLists::get_ids(size_t list_no) const { assert(list_no < nlist); return ids[list_no].data(); } @@ -267,7 +266,7 @@ void HStackInvertedLists::release_codes(size_t, const uint8_t* codes) const { delete[] codes; } -const Index::idx_t* HStackInvertedLists::get_ids(size_t list_no) const { +const idx_t* HStackInvertedLists::get_ids(size_t list_no) const { idx_t *ids = new idx_t[list_size(list_no)], *c = ids; for (int i = 0; i < ils.size(); i++) { @@ -281,8 +280,7 @@ const Index::idx_t* HStackInvertedLists::get_ids(size_t list_no) const { return ids; } -Index::idx_t HStackInvertedLists::get_single_id(size_t list_no, size_t offset) - const { +idx_t HStackInvertedLists::get_single_id(size_t list_no, size_t offset) const { for (int i = 0; i < ils.size(); i++) { const InvertedLists* il = ils[i]; size_t sz = il->list_size(list_no); @@ -312,8 +310,6 @@ void HStackInvertedLists::prefetch_lists(const idx_t* list_nos, int nlist) namespace { -using idx_t = InvertedLists::idx_t; - idx_t translate_list_no(const SliceInvertedLists* sil, idx_t list_no) { FAISS_THROW_IF_NOT(list_no >= 0 && list_no < sil->nlist); return list_no + sil->i0; @@ -349,12 +345,11 @@ void SliceInvertedLists::release_codes(size_t list_no, const uint8_t* codes) return il->release_codes(translate_list_no(this, list_no), codes); } -const Index::idx_t* SliceInvertedLists::get_ids(size_t list_no) const { +const idx_t* SliceInvertedLists::get_ids(size_t list_no) const { return il->get_ids(translate_list_no(this, list_no)); } -Index::idx_t SliceInvertedLists::get_single_id(size_t list_no, size_t offset) - const { +idx_t SliceInvertedLists::get_single_id(size_t list_no, size_t offset) const { return il->get_single_id(translate_list_no(this, list_no), offset); } @@ -380,8 +375,6 @@ void SliceInvertedLists::prefetch_lists(const idx_t* list_nos, int nlist) namespace { -using idx_t = InvertedLists::idx_t; - // find the invlist this number belongs to int translate_list_no(const VStackInvertedLists* vil, idx_t list_no) { FAISS_THROW_IF_NOT(list_no >= 0 && list_no < vil->nlist); @@ -449,14 +442,13 @@ void VStackInvertedLists::release_codes(size_t list_no, const uint8_t* codes) return ils[i]->release_codes(list_no, codes); } -const Index::idx_t* VStackInvertedLists::get_ids(size_t list_no) const { +const idx_t* VStackInvertedLists::get_ids(size_t list_no) const { int i = translate_list_no(this, list_no); list_no -= cumsz[i]; return ils[i]->get_ids(list_no); } -Index::idx_t VStackInvertedLists::get_single_id(size_t list_no, size_t offset) - const { +idx_t VStackInvertedLists::get_single_id(size_t list_no, size_t offset) const { int i = translate_list_no(this, list_no); list_no -= cumsz[i]; return ils[i]->get_single_id(list_no, offset); diff --git a/faiss/invlists/InvertedLists.h b/faiss/invlists/InvertedLists.h index 80c084b09..94a78fe76 100644 --- a/faiss/invlists/InvertedLists.h +++ b/faiss/invlists/InvertedLists.h @@ -15,7 +15,7 @@ * the interface. */ -#include +#include #include namespace faiss { @@ -28,8 +28,6 @@ namespace faiss { * are allowed */ struct InvertedLists { - typedef Index::idx_t idx_t; - size_t nlist; ///< number of possible key values size_t code_size; ///< code size per vector in bytes diff --git a/faiss/invlists/OnDiskInvertedLists.cpp b/faiss/invlists/OnDiskInvertedLists.cpp index 5027c0e86..825ccfbb9 100644 --- a/faiss/invlists/OnDiskInvertedLists.cpp +++ b/faiss/invlists/OnDiskInvertedLists.cpp @@ -154,7 +154,7 @@ struct OnDiskInvertedLists::OngoingPrefetch { const OnDiskInvertedLists* od = pf->od; od->locks->lock_1(list_no); size_t n = od->list_size(list_no); - const Index::idx_t* idx = od->get_ids(list_no); + const idx_t* idx = od->get_ids(list_no); const uint8_t* codes = od->get_codes(list_no); int cs = 0; for (size_t i = 0; i < n; i++) { @@ -389,7 +389,7 @@ const uint8_t* OnDiskInvertedLists::get_codes(size_t list_no) const { return ptr + lists[list_no].offset; } -const Index::idx_t* OnDiskInvertedLists::get_ids(size_t list_no) const { +const idx_t* OnDiskInvertedLists::get_ids(size_t list_no) const { if (lists[list_no].offset == INVALID_OFFSET) { return nullptr; } @@ -781,7 +781,7 @@ InvertedLists* OnDiskInvertedListsIOHook::read_ArrayInvertedLists( OnDiskInvertedLists::List& l = ails->lists[i]; l.size = l.capacity = sizes[i]; l.offset = o; - o += l.size * (sizeof(OnDiskInvertedLists::idx_t) + ails->code_size); + o += l.size * (sizeof(idx_t) + ails->code_size); } // resume normal reading of file fseek(fdesc, o, SEEK_SET); diff --git a/faiss/python/python_callbacks.cpp b/faiss/python/python_callbacks.cpp index a5379935a..c99c1a3f7 100644 --- a/faiss/python/python_callbacks.cpp +++ b/faiss/python/python_callbacks.cpp @@ -118,7 +118,7 @@ PyCallbackIDSelector::PyCallbackIDSelector(PyObject* callback) Py_INCREF(callback); } -bool PyCallbackIDSelector::is_member(idx_t id) const { +bool PyCallbackIDSelector::is_member(faiss::idx_t id) const { FAISS_THROW_IF_NOT((id >> 32) == 0); PyThreadLock gil; PyObject* result = PyObject_CallFunction(callback, "(n)", int(id)); diff --git a/faiss/python/python_callbacks.h b/faiss/python/python_callbacks.h index 605e76c41..421239bd0 100644 --- a/faiss/python/python_callbacks.h +++ b/faiss/python/python_callbacks.h @@ -54,7 +54,7 @@ struct PyCallbackIDSelector : faiss::IDSelector { explicit PyCallbackIDSelector(PyObject* callback); - bool is_member(idx_t id) const override; + bool is_member(faiss::idx_t id) const override; ~PyCallbackIDSelector() override; }; diff --git a/faiss/python/swigfaiss.swig b/faiss/python/swigfaiss.swig index f0bcb3047..e9a40428a 100644 --- a/faiss/python/swigfaiss.swig +++ b/faiss/python/swigfaiss.swig @@ -1006,8 +1006,8 @@ float * cast_integer_to_float_ptr (int64_t x) { return (float*)x; } -faiss::Index::idx_t* cast_integer_to_idx_t_ptr (int64_t x) { - return (faiss::Index::idx_t*)x; +faiss::idx_t* cast_integer_to_idx_t_ptr (int64_t x) { + return (faiss::idx_t*)x; } int * cast_integer_to_int_ptr (int64_t x) { diff --git a/faiss/utils/extra_distances.cpp b/faiss/utils/extra_distances.cpp index a0e2dfc51..dc60611a9 100644 --- a/faiss/utils/extra_distances.cpp +++ b/faiss/utils/extra_distances.cpp @@ -92,7 +92,7 @@ void knn_extra_metrics_template( template struct ExtraDistanceComputer : FlatCodesDistanceComputer { VD vd; - Index::idx_t nb; + idx_t nb; const float* q; const float* b; diff --git a/tests/test_binary_flat.cpp b/tests/test_binary_flat.cpp index c974d0bdd..048edce20 100644 --- a/tests/test_binary_flat.cpp +++ b/tests/test_binary_flat.cpp @@ -42,7 +42,7 @@ TEST(BinaryFlat, accuracy) { } int k = 5; - std::vector nns(k * nq); + std::vector nns(k * nq); std::vector dis(k * nq); index.search(nq, queries.data(), k, dis.data(), nns.data()); diff --git a/tests/test_dealloc_invlists.cpp b/tests/test_dealloc_invlists.cpp index 528869c7f..f1b69379a 100644 --- a/tests/test_dealloc_invlists.cpp +++ b/tests/test_dealloc_invlists.cpp @@ -24,8 +24,6 @@ using namespace faiss; namespace { -typedef Index::idx_t idx_t; - // dimension of the vectors to index int d = 32; diff --git a/tests/test_io.py b/tests/test_io.py index 7b0fc2f09..8ba2d13b4 100644 --- a/tests/test_io.py +++ b/tests/test_io.py @@ -11,7 +11,7 @@ import os import io import sys import pickle -from multiprocessing.dummy import Pool as ThreadPool +from multiprocessing.pool import ThreadPool class TestIOVariants(unittest.TestCase): diff --git a/tests/test_ivfpq_codec.cpp b/tests/test_ivfpq_codec.cpp index 3044c77ef..9ea5ce853 100644 --- a/tests/test_ivfpq_codec.cpp +++ b/tests/test_ivfpq_codec.cpp @@ -34,7 +34,7 @@ double eval_codec_error(long ncentroids, long m, const std::vector& v) { // encode and decode to compute reconstruction error - std::vector keys(nb); + std::vector keys(nb); std::vector codes(nb * m); index.encode_multiple(nb, keys.data(), v.data(), codes.data(), true); diff --git a/tests/test_ivfpq_indexing.cpp b/tests/test_ivfpq_indexing.cpp index d46b2ed21..334256096 100644 --- a/tests/test_ivfpq_indexing.cpp +++ b/tests/test_ivfpq_indexing.cpp @@ -71,14 +71,14 @@ TEST(IVFPQ, accuracy) { queries[i] = distrib(rng); } - std::vector gt_nns(nq); + std::vector gt_nns(nq); std::vector gt_dis(nq); index_gt.search(nq, queries.data(), 1, gt_dis.data(), gt_nns.data()); index.nprobe = 5; int k = 5; - std::vector nns(k * nq); + std::vector nns(k * nq); std::vector dis(k * nq); index.search(nq, queries.data(), k, dis.data(), nns.data()); diff --git a/tests/test_lowlevel_ivf.cpp b/tests/test_lowlevel_ivf.cpp index 44450fcac..e28e2a946 100644 --- a/tests/test_lowlevel_ivf.cpp +++ b/tests/test_lowlevel_ivf.cpp @@ -29,8 +29,6 @@ using namespace faiss; namespace { -typedef Index::idx_t idx_t; - // dimension of the vectors to index int d = 32; diff --git a/tests/test_mem_leak.cpp b/tests/test_mem_leak.cpp index b4bb40ada..b0232a85a 100644 --- a/tests/test_mem_leak.cpp +++ b/tests/test_mem_leak.cpp @@ -40,7 +40,7 @@ TEST(MEM_LEAK, ivfflat) { double t0 = getmillisecs(); for (int i = 0; i < N2; i++) { - std::vector I(10 * bs); + std::vector I(10 * bs); std::vector D(10 * bs); tfidf_faiss_index.search( diff --git a/tests/test_merge.cpp b/tests/test_merge.cpp index f39b23b8d..493fb3ccd 100644 --- a/tests/test_merge.cpp +++ b/tests/test_merge.cpp @@ -48,7 +48,7 @@ struct Tempfilename { pthread_mutex_t Tempfilename::mutex = PTHREAD_MUTEX_INITIALIZER; -typedef faiss::Index::idx_t idx_t; +typedef faiss::idx_t idx_t; // parameters to use for the test int d = 64; diff --git a/tests/test_ondisk_ivf.cpp b/tests/test_ondisk_ivf.cpp index 6ff383e5c..94c23381e 100644 --- a/tests/test_ondisk_ivf.cpp +++ b/tests/test_ondisk_ivf.cpp @@ -80,10 +80,10 @@ TEST(ONDISK, make_invlists) { int ntot = 0; for (int i = 0; i < nlist; i++) { int size = ivf.list_size(i); - const faiss::Index::idx_t* ids = ivf.get_ids(i); + const faiss::idx_t* ids = ivf.get_ids(i); const uint8_t* codes = ivf.get_codes(i); for (int j = 0; j < size; j++) { - faiss::Index::idx_t id = ids[j]; + faiss::idx_t id = ids[j]; const int* ar = (const int*)&codes[code_size * j]; EXPECT_EQ(ar[0], id); EXPECT_EQ(ar[1], i); @@ -113,7 +113,7 @@ TEST(ONDISK, test_add) { faiss::float_rand(xq.data(), d * nq, 34567); std::vector ref_D(nq * k); - std::vector ref_I(nq * k); + std::vector ref_I(nq * k); index.search(nq, xq.data(), k, ref_D.data(), ref_I.data()); @@ -131,7 +131,7 @@ TEST(ONDISK, test_add) { index2.add(nb, xb.data()); std::vector new_D(nq * k); - std::vector new_I(nq * k); + std::vector new_I(nq * k); index2.search(nq, xq.data(), k, new_D.data(), new_I.data()); @@ -146,7 +146,7 @@ TEST(ONDISK, test_add) { faiss::Index* index3 = faiss::read_index(filename2.c_str()); std::vector new_D(nq * k); - std::vector new_I(nq * k); + std::vector new_I(nq * k); index3->search(nq, xq.data(), k, new_D.data(), new_I.data()); @@ -192,10 +192,10 @@ TEST(ONDISK, make_invlists_threaded) { int ntot = 0; for (int i = 0; i < nlist; i++) { int size = ivf.list_size(i); - const faiss::Index::idx_t* ids = ivf.get_ids(i); + const faiss::idx_t* ids = ivf.get_ids(i); const uint8_t* codes = ivf.get_codes(i); for (int j = 0; j < size; j++) { - faiss::Index::idx_t id = ids[j]; + faiss::idx_t id = ids[j]; const int* ar = (const int*)&codes[code_size * j]; EXPECT_EQ(ar[0], id); EXPECT_EQ(ar[1], i); diff --git a/tests/test_pairs_decoding.cpp b/tests/test_pairs_decoding.cpp index daaa0ff31..5c7921fbc 100644 --- a/tests/test_pairs_decoding.cpp +++ b/tests/test_pairs_decoding.cpp @@ -21,7 +21,7 @@ namespace { -typedef faiss::Index::idx_t idx_t; +typedef faiss::idx_t idx_t; /************************************************************* * Test utils diff --git a/tests/test_params_override.cpp b/tests/test_params_override.cpp index 809bc10d0..8891d0e55 100644 --- a/tests/test_params_override.cpp +++ b/tests/test_params_override.cpp @@ -27,8 +27,6 @@ using namespace faiss; namespace { -typedef Index::idx_t idx_t; - // dimension of the vectors to index int d = 32; diff --git a/tests/test_sliding_ivf.cpp b/tests/test_sliding_ivf.cpp index 851713c52..ea9e53d6b 100644 --- a/tests/test_sliding_ivf.cpp +++ b/tests/test_sliding_ivf.cpp @@ -22,8 +22,6 @@ using namespace faiss; -typedef Index::idx_t idx_t; - // dimension of the vectors to index int d = 32; @@ -81,7 +79,7 @@ void make_index_slices( Index* index = sub_indexes.back().get(); auto xb = make_data(nb * d); - std::vector ids(nb); + std::vector ids(nb); std::mt19937 rng; std::uniform_int_distribution<> distrib; for (int j = 0; j < nb; j++) { diff --git a/tests/test_threaded_index.cpp b/tests/test_threaded_index.cpp index 307ff685b..b10d3806e 100644 --- a/tests/test_threaded_index.cpp +++ b/tests/test_threaded_index.cpp @@ -19,6 +19,8 @@ namespace { struct TestException : public std::exception {}; +using idx_t = faiss::idx_t; + struct MockIndex : public faiss::Index { explicit MockIndex(idx_t d) : faiss::Index(d) { resetMock(); @@ -66,7 +68,7 @@ struct MockIndex : public faiss::Index { template struct MockThreadedIndex : public faiss::ThreadedIndex { - using idx_t = faiss::Index::idx_t; + using idx_t = faiss::idx_t; explicit MockThreadedIndex(bool threaded) : faiss::ThreadedIndex(threaded) {} @@ -178,7 +180,7 @@ TEST(ThreadedIndex, TestReplica) { std::vector x(n * d); std::vector distances(n * k); - std::vector labels(n * k); + std::vector labels(n * k); replica.add(n, x.data()); @@ -227,7 +229,7 @@ TEST(ThreadedIndex, TestShards) { std::vector x(n * d); std::vector distances(n * k); - std::vector labels(n * k); + std::vector labels(n * k); shards.add(n, x.data()); diff --git a/tests/test_transfer_invlists.cpp b/tests/test_transfer_invlists.cpp index c0a9f67ec..309a331dd 100644 --- a/tests/test_transfer_invlists.cpp +++ b/tests/test_transfer_invlists.cpp @@ -33,7 +33,7 @@ int nlist = 40; using namespace faiss; -typedef faiss::Index::idx_t idx_t; +typedef faiss::idx_t idx_t; std::vector get_data(size_t nb, int seed) { std::vector x(nb * d); diff --git a/tutorial/cpp/1-Flat.cpp b/tutorial/cpp/1-Flat.cpp index edd1426be..819e41957 100644 --- a/tutorial/cpp/1-Flat.cpp +++ b/tutorial/cpp/1-Flat.cpp @@ -12,7 +12,7 @@ #include // 64-bit int -using idx_t = faiss::Index::idx_t; +using idx_t = faiss::idx_t; int main() { int d = 64; // dimension diff --git a/tutorial/cpp/2-IVFFlat.cpp b/tutorial/cpp/2-IVFFlat.cpp index e107c8fc6..febd5be04 100644 --- a/tutorial/cpp/2-IVFFlat.cpp +++ b/tutorial/cpp/2-IVFFlat.cpp @@ -13,7 +13,7 @@ #include #include -using idx_t = faiss::Index::idx_t; +using idx_t = faiss::idx_t; int main() { int d = 64; // dimension diff --git a/tutorial/cpp/3-IVFPQ.cpp b/tutorial/cpp/3-IVFPQ.cpp index ba1bc5b4c..c84e52e86 100644 --- a/tutorial/cpp/3-IVFPQ.cpp +++ b/tutorial/cpp/3-IVFPQ.cpp @@ -12,7 +12,7 @@ #include #include -using idx_t = faiss::Index::idx_t; +using idx_t = faiss::idx_t; int main() { int d = 64; // dimension