Summary: Pull Request resolved: https://github.com/facebookresearch/faiss/pull/5696 `compute_inner_product<SIMDLevel::AVX2>` scores 9-bit RaBitQ codes (`ex_bits == 8`) in groups of 8 dimensions. Before this change, each group loaded one sign byte and broadcast it. The broadcast uses the shuffle port, which limits the loop. This change makes these changes to the loop: - The loop loads one 32-bit sign word and broadcasts it once for each 32 dimensions. - Each group of 8 dimensions ANDs the word with `positions << 8k` and compares the result with `cmpeq`. A `cmpgt` against zero is not correct here, because bit 31 of the shifted positions is a sign bit. - The loop uses two accumulators. - An 8-dimension loop handles the remaining full groups. `ip_scalar` handles the tail, as before. The two accumulators change the float summation order. The benchmark checksums agree to 6 significant digits. AI explanation of why: ``` What limits a loop like this A loop runs no faster than its tightest limit. This loop has three possible limits: 1. Micro-op count. The core issues about 4 micro-ops per cycle. 2. Port 5, the shuffle port. Broadcasts and byte-to-int expansions can run only there. 3. The FMA chain. With one accumulator, each FMA must wait for the previous one. Each FMA takes 4 cycles, so one group of 8 dims takes at least 4 cycles. Measured values, per group of 8 dims (d=768, codes in L2) ┌────────────────────────────────────────┬────────┬──────────────┬─────────────┬─────────────────┐ │ loop │ cycles │ instructions │ port 5 µops │ FMA chain floor │ ├────────────────────────────────────────┼────────┼──────────────┼─────────────┼─────────────────┤ │ V1 (contributor), 1 accumulator │ 5.53 │ 18.4 │ 4.12 │ 4.0 │ ├────────────────────────────────────────┼────────┼──────────────┼─────────────┼─────────────────┤ │ V1, 2 accumulators │ 5.63 │ 15.4 │ 4.78 │ 2.0 │ ├────────────────────────────────────────┼────────┼──────────────┼─────────────┼─────────────────┤ │ sign word, 1 accumulator │ 4.43 │ 10.4 │ 2.95 │ 4.0 │ ├────────────────────────────────────────┼────────┼──────────────┼─────────────┼─────────────────┤ │ sign word, 2 accumulators (D122893433) │ 3.77 │ 10.5 │ 2.99 │ 2.0 │ └────────────────────────────────────────┴────────┴──────────────┴─────────────┴─────────────────┘ With 4,096 codes, which do not fit in L2, the values are 6.20, 5.96, 5.15 and 4.66 cycles. The new loop is 1.33x faster than V1 there, which agrees with the faiss benchmark (1.33x–1.36x). Step 1: the sign-word broadcast removes micro-ops In V1, each group of 8 dims gets its sign byte in four instructions: movzx eax, BYTE PTR [sign + i/8] ; load one byte vmovd xmm1, eax ; copy it into a vector register (port 5) vpbroadcastd ymm1, xmm1 ; copy it into all 8 lanes (port 5) The loop also does its index and loop-control work (shr, mov, add, cmp, jae) once per 8 dims. In total, each group needs 18 instructions. The new loop loads a 32-bit sign word once per 32 dims. The compiler turns the memcpy plus _mm256_set1_epi32 into one instruction: vpbroadcastd ymm0, DWORD PTR [sign + i/8] A broadcast that reads directly from memory runs on the load port only. It does not use port 5. Each of the four groups then needs only vpand and vpcmpeqd against its shifted position constant. The compiler loads these constants before the loop, so they cost nothing per pass. The loop-control work also runs once per 32 dims, not once per 8. Result: instructions per group go from 18.4 to 10.4, and port 5 micro-ops from 4.12 to 2.95. Cycles go from 5.53 to 4.43, which is 1.25x. Step 2: the second accumulator shortens the FMA chain After step 1, the micro-op limit is about 2.6 cycles and the port 5 limit is about 3.0 cycles. But one accumulator still forces 4 cycles per group. The measured 4.43 cycles is close to that floor, so the FMA chain is now the limit. With two accumulators, two chains run at the same time, and the floor falls to 2.0 cycles. The next limit is port 5 at about 3 cycles. Cycles go from 4.43 to 3.77, which is 1.18x. Why the second accumulator alone did not help V1 V1 needs 5.53 cycles because of its micro-op count and port 5, not because of the FMA chain: 5.53 is well above the 4-cycle floor. A second accumulator removes a limit that V1 does not reach. My earlier standalone test found the same result. Thus each change helps only together with the other: - The broadcast change brings the loop down to the FMA floor. - The second accumulator then lowers the floor. The cmpeq change cmpeq is not faster than cmpgt; both cost the same. It is necessary for correctness: the group that tests bits 24–31 uses 128 << 24 = 0x80000000. As a signed integer this value is negative, so cmpgt(x, 0) would give the wrong result for that lane. ``` Reviewed By: junjieqi Differential Revision: D122893433 fbshipit-source-id: e6c42d33e138614589e8f80c059df3652a7ddfcd
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 Meta's Fundamental AI Research group.
News
See CHANGELOG.md for detailed information about latest features.
Introduction
Faiss contains several methods for similarity search. It assumes that the instances are represented as vectors and are identified by an integer, and that the vectors can be compared with L2 (Euclidean) distances or dot products. Vectors that are similar to a query vector are those that have the lowest L2 distance or the highest dot product with the query vector. It also supports cosine similarity, since this is a dot product on normalized vectors.
Some of the methods, like those based on binary vectors and compact quantization codes, solely use a compressed representation of the vectors and do not require to keep the original vectors. This generally comes at the cost of a less precise search but these methods can scale to billions of vectors in main memory on a single server. Other methods, like HNSW and NSG add an indexing structure on top of the raw vectors to make searching more efficient.
The GPU implementation can accept input from either CPU or GPU memory. On a server with GPUs, the GPU indexes can be used a drop-in replacement for the CPU indexes (e.g., replace IndexFlatL2 with GpuIndexFlatL2) and copies to/from GPU memory are handled automatically. Results will be faster however if both input and output remain resident on the GPU. Both single and multi-GPU usage is supported.
Installing
Faiss comes with precompiled libraries for Anaconda in Python, see faiss-cpu, faiss-gpu and faiss-gpu-cuvs. The library is mostly implemented in C++, the only dependency is a BLAS implementation. Optional GPU support is provided via CUDA or AMD ROCm, and the Python interface is also optional. The backend GPU implementations of NVIDIA cuVS can also be enabled optionally. It compiles with cmake. See INSTALL.md for details.
How Faiss works
Faiss is built around an index type that stores a set of vectors, and provides a function to search in them with L2 and/or dot product vector comparison. Some index types are simple baselines, such as exact search. Most of the available indexing structures correspond to various trade-offs with respect to
- search time
- search quality
- memory used per index vector
- training time
- adding time
- need for external data for unsupervised training
The optional GPU implementation provides what is likely (as of March 2017) the fastest exact and approximate (compressed-domain) nearest neighbor search implementation for high-dimensional vectors, fastest Lloyd's k-means, and fastest small k-selection algorithm known. The implementation is detailed here.
Full documentation of Faiss
The following are entry points for documentation:
- the full documentation can be found on the wiki page, including a tutorial, a FAQ and a troubleshooting section
- the doxygen documentation gives per-class information extracted from code comments
- to reproduce results from our research papers, Polysemous codes and Billion-scale similarity search with GPUs, refer to the benchmarks README. For Link and code: Fast indexing with graphs and compact regression codes, see the link_and_code README
Authors
The main authors of Faiss are:
- Hervé Jégou initiated the Faiss project and wrote its first implementation
- Matthijs Douze implemented most of the CPU Faiss
- Jeff Johnson implemented all of the GPU Faiss
- Lucas Hosseini implemented the binary indexes and the build system
- Chengqi Deng implemented NSG, NNdescent and much of the additive quantization code.
- Alexandr Guzhva many optimizations: SIMD, memory allocation and layout, fast decoding kernels for vector codecs, etc.
- Gergely Szilvasy build system, benchmarking framework.
Reference
References to cite when you use Faiss in a research paper:
@article{douze2024faiss,
title={The Faiss library},
author={Matthijs Douze and Alexandr Guzhva and Chengqi Deng and Jeff Johnson and Gergely Szilvasy and Pierre-Emmanuel Mazaré and Maria Lomeli and Lucas Hosseini and Hervé Jégou},
year={2024},
eprint={2401.08281},
archivePrefix={arXiv},
primaryClass={cs.LG}
}
For the GPU version of Faiss, please cite:
@article{johnson2019billion,
title={Billion-scale similarity search with {GPUs}},
author={Johnson, Jeff and Douze, Matthijs and J{\'e}gou, Herv{\'e}},
journal={IEEE Transactions on Big Data},
volume={7},
number={3},
pages={535--547},
year={2019},
publisher={IEEE}
}
Join the Faiss community
For public discussion of Faiss or for questions, visit https://github.com/facebookresearch/faiss/discussions.
We monitor the issues page of the repository. You can report bugs, ask questions, etc.
Legal
Faiss is MIT-licensed, refer to the LICENSE file in the top level directory.
Copyright © Meta Platforms, Inc.