Skip to content

Commit 61819c5

Browse files
bkarsinbkarsin
committed
clang-format
1 parent 2b41a0e commit 61819c5

8 files changed

Lines changed: 88 additions & 83 deletions

File tree

‎cpp/src/neighbors/detail/vamana/greedy_search.cuh‎

Lines changed: 3 additions & 4 deletions
Original file line numberDiff line numberDiff line change
@@ -103,7 +103,6 @@ __global__ void GreedySearchKernel(
103103

104104
static __shared__ Point<T, accT> s_query;
105105

106-
107106
union ShmemLayout {
108107
// All blocksort sizes have same alignment (16)
109108
typename cub::BlockMergeSort<DistPair<IdxT, accT>, 32, 1>::TempStorage sort_mem;
@@ -113,15 +112,15 @@ __global__ void GreedySearchKernel(
113112
DistPair<IdxT, accT> candidate_queue;
114113
};
115114

116-
int align_padding = (((dim-1)/alignof(ShmemLayout))+1)*alignof(ShmemLayout) - dim;
115+
int align_padding = (((dim - 1) / alignof(ShmemLayout)) + 1) * alignof(ShmemLayout) - dim;
117116

118117
// Dynamic shared memory used for blocksort, temp vector storage, and neighborhood list
119118
extern __shared__ __align__(alignof(ShmemLayout)) char smem[];
120119

121120
size_t smem_offset = sort_smem_size; // temp sorting memory takes first chunk
122121

123122
T* s_coords = reinterpret_cast<T*>(&smem[smem_offset]);
124-
smem_offset += (dim+align_padding) * sizeof(T);
123+
smem_offset += (dim + align_padding) * sizeof(T);
125124

126125
Node<accT>* topk_pq = reinterpret_cast<Node<accT>*>(&smem[smem_offset]);
127126
smem_offset += topk * sizeof(Node<accT>);
@@ -173,7 +172,7 @@ __global__ void GreedySearchKernel(
173172

174173
if (threadIdx.x == 0) { heap_queue.insert_back(medoid_dist, medoid_id); }
175174
__syncthreads();
176-
175+
177176
while (cand_q_size != 0) {
178177
__syncthreads();
179178

‎cpp/src/neighbors/detail/vamana/robust_prune.cuh‎

Lines changed: 4 additions & 4 deletions
Original file line numberDiff line numberDiff line change
@@ -145,11 +145,11 @@ __global__ void RobustPruneKernel(
145145
// Dynamic shared memory used for blocksort, temp vector storage, and neighborhood list
146146
extern __shared__ __align__(alignof(ShmemLayout)) char smem[];
147147

148-
int align_padding = (((dim-1)/alignof(ShmemLayout))+1)*alignof(ShmemLayout) - dim;
148+
int align_padding = (((dim - 1) / alignof(ShmemLayout)) + 1) * alignof(ShmemLayout) - dim;
149149

150-
T* s_coords = reinterpret_cast<T*>(&smem[sort_smem_size]);
151-
DistPair<IdxT, accT>* new_nbh_list =
152-
reinterpret_cast<DistPair<IdxT, accT>*>(&smem[(dim+align_padding) * sizeof(T) + sort_smem_size]);
150+
T* s_coords = reinterpret_cast<T*>(&smem[sort_smem_size]);
151+
DistPair<IdxT, accT>* new_nbh_list = reinterpret_cast<DistPair<IdxT, accT>*>(
152+
&smem[(dim + align_padding) * sizeof(T) + sort_smem_size]);
153153

154154
static __shared__ Point<T, accT> s_query;
155155
s_query.coords = s_coords;

‎cpp/src/neighbors/vamana.cuh‎

Lines changed: 1 addition & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -85,8 +85,7 @@ index<T, IdxT> build(
8585
const index_params& params,
8686
raft::mdspan<const T, raft::matrix_extent<int64_t>, raft::row_major, Accessor> dataset)
8787
{
88-
return cuvs::neighbors::vamana::detail::build<T, IdxT, Accessor>(
89-
res, params, dataset);
88+
return cuvs::neighbors::vamana::detail::build<T, IdxT, Accessor>(res, params, dataset);
9089
}
9190

9291
template <typename T, typename IdxT>

‎cpp/src/neighbors/vamana_build_float.cu‎

Lines changed: 15 additions & 15 deletions
Original file line numberDiff line numberDiff line change
@@ -19,21 +19,21 @@
1919

2020
namespace cuvs::neighbors::vamana {
2121

22-
#define RAFT_INST_VAMANA_BUILD(T, IdxT) \
23-
auto build(raft::resources const& handle, \
24-
const cuvs::neighbors::vamana::index_params& params, \
25-
raft::device_matrix_view<const T, int64_t, raft::row_major> dataset) \
26-
->cuvs::neighbors::vamana::index<T, IdxT> \
27-
{ \
28-
return cuvs::neighbors::vamana::build<T, IdxT>(handle, params, dataset); \
29-
} \
30-
\
31-
auto build(raft::resources const& handle, \
32-
const cuvs::neighbors::vamana::index_params& params, \
33-
raft::host_matrix_view<const T, int64_t, raft::row_major> dataset) \
34-
->cuvs::neighbors::vamana::index<T, IdxT> \
35-
{ \
36-
return cuvs::neighbors::vamana::build<T, IdxT>(handle, params, dataset); \
22+
#define RAFT_INST_VAMANA_BUILD(T, IdxT) \
23+
auto build(raft::resources const& handle, \
24+
const cuvs::neighbors::vamana::index_params& params, \
25+
raft::device_matrix_view<const T, int64_t, raft::row_major> dataset) \
26+
->cuvs::neighbors::vamana::index<T, IdxT> \
27+
{ \
28+
return cuvs::neighbors::vamana::build<T, IdxT>(handle, params, dataset); \
29+
} \
30+
\
31+
auto build(raft::resources const& handle, \
32+
const cuvs::neighbors::vamana::index_params& params, \
33+
raft::host_matrix_view<const T, int64_t, raft::row_major> dataset) \
34+
->cuvs::neighbors::vamana::index<T, IdxT> \
35+
{ \
36+
return cuvs::neighbors::vamana::build<T, IdxT>(handle, params, dataset); \
3737
}
3838

3939
RAFT_INST_VAMANA_BUILD(float, uint32_t);

‎cpp/src/neighbors/vamana_build_int8.cu‎

Lines changed: 15 additions & 15 deletions
Original file line numberDiff line numberDiff line change
@@ -19,21 +19,21 @@
1919

2020
namespace cuvs::neighbors::vamana {
2121

22-
#define RAFT_INST_VAMANA_BUILD(T, IdxT) \
23-
auto build(raft::resources const& handle, \
24-
const cuvs::neighbors::vamana::index_params& params, \
25-
raft::device_matrix_view<const T, int64_t, raft::row_major> dataset) \
26-
->cuvs::neighbors::vamana::index<T, IdxT> \
27-
{ \
28-
return cuvs::neighbors::vamana::build<T, IdxT>(handle, params, dataset); \
29-
} \
30-
\
31-
auto build(raft::resources const& handle, \
32-
const cuvs::neighbors::vamana::index_params& params, \
33-
raft::host_matrix_view<const T, int64_t, raft::row_major> dataset) \
34-
->cuvs::neighbors::vamana::index<T, IdxT> \
35-
{ \
36-
return cuvs::neighbors::vamana::build<T, IdxT>(handle, params, dataset); \
22+
#define RAFT_INST_VAMANA_BUILD(T, IdxT) \
23+
auto build(raft::resources const& handle, \
24+
const cuvs::neighbors::vamana::index_params& params, \
25+
raft::device_matrix_view<const T, int64_t, raft::row_major> dataset) \
26+
->cuvs::neighbors::vamana::index<T, IdxT> \
27+
{ \
28+
return cuvs::neighbors::vamana::build<T, IdxT>(handle, params, dataset); \
29+
} \
30+
\
31+
auto build(raft::resources const& handle, \
32+
const cuvs::neighbors::vamana::index_params& params, \
33+
raft::host_matrix_view<const T, int64_t, raft::row_major> dataset) \
34+
->cuvs::neighbors::vamana::index<T, IdxT> \
35+
{ \
36+
return cuvs::neighbors::vamana::build<T, IdxT>(handle, params, dataset); \
3737
}
3838

3939
RAFT_INST_VAMANA_BUILD(int8_t, uint32_t);

‎cpp/src/neighbors/vamana_build_uint8.cu‎

Lines changed: 15 additions & 15 deletions
Original file line numberDiff line numberDiff line change
@@ -19,21 +19,21 @@
1919

2020
namespace cuvs::neighbors::vamana {
2121

22-
#define RAFT_INST_VAMANA_BUILD(T, IdxT) \
23-
auto build(raft::resources const& handle, \
24-
const cuvs::neighbors::vamana::index_params& params, \
25-
raft::device_matrix_view<const T, int64_t, raft::row_major> dataset) \
26-
->cuvs::neighbors::vamana::index<T, IdxT> \
27-
{ \
28-
return cuvs::neighbors::vamana::build<T, IdxT>(handle, params, dataset); \
29-
} \
30-
\
31-
auto build(raft::resources const& handle, \
32-
const cuvs::neighbors::vamana::index_params& params, \
33-
raft::host_matrix_view<const T, int64_t, raft::row_major> dataset) \
34-
->cuvs::neighbors::vamana::index<T, IdxT> \
35-
{ \
36-
return cuvs::neighbors::vamana::build<T, IdxT>(handle, params, dataset); \
22+
#define RAFT_INST_VAMANA_BUILD(T, IdxT) \
23+
auto build(raft::resources const& handle, \
24+
const cuvs::neighbors::vamana::index_params& params, \
25+
raft::device_matrix_view<const T, int64_t, raft::row_major> dataset) \
26+
->cuvs::neighbors::vamana::index<T, IdxT> \
27+
{ \
28+
return cuvs::neighbors::vamana::build<T, IdxT>(handle, params, dataset); \
29+
} \
30+
\
31+
auto build(raft::resources const& handle, \
32+
const cuvs::neighbors::vamana::index_params& params, \
33+
raft::host_matrix_view<const T, int64_t, raft::row_major> dataset) \
34+
->cuvs::neighbors::vamana::index<T, IdxT> \
35+
{ \
36+
return cuvs::neighbors::vamana::build<T, IdxT>(handle, params, dataset); \
3737
}
3838

3939
RAFT_INST_VAMANA_BUILD(uint8_t, uint32_t);

‎cpp/src/neighbors/vamana_serialize.cuh‎

Lines changed: 2 additions & 3 deletions
Original file line numberDiff line numberDiff line change
@@ -28,10 +28,9 @@ namespace cuvs::neighbors::vamana {
2828
#define CUVS_INST_VAMANA_SERIALIZE(DTYPE) \
2929
void serialize(raft::resources const& handle, \
3030
const std::string& file_prefix, \
31-
const cuvs::neighbors::vamana::index<DTYPE, uint32_t>& index_) \
31+
const cuvs::neighbors::vamana::index<DTYPE, uint32_t>& index_) \
3232
{ \
33-
cuvs::neighbors::vamana::detail::serialize<DTYPE, uint32_t>( \
34-
handle, file_prefix, index_); \
33+
cuvs::neighbors::vamana::detail::serialize<DTYPE, uint32_t>(handle, file_prefix, index_); \
3534
};
3635

3736
/** @} */ // end group vamana

‎examples/cpp/src/vamana_example.cu‎

Lines changed: 33 additions & 25 deletions
Original file line numberDiff line numberDiff line change
@@ -29,10 +29,14 @@
2929
#include "common.cuh"
3030

3131
template <typename T>
32-
void vamana_build_and_write(raft::device_resources const &dev_resources,
32+
void vamana_build_and_write(raft::device_resources const& dev_resources,
3333
raft::device_matrix_view<const T, int64_t> dataset,
34-
std::string out_fname, int degree, int visited_size,
35-
float max_fraction, int iters) {
34+
std::string out_fname,
35+
int degree,
36+
int visited_size,
37+
float max_fraction,
38+
int iters)
39+
{
3640
using namespace cuvs::neighbors;
3741

3842
// use default index parameters
@@ -46,23 +50,24 @@ void vamana_build_and_write(raft::device_resources const &dev_resources,
4650

4751
auto start = std::chrono::system_clock::now();
4852
auto index = vamana::build(dev_resources, index_params, dataset);
49-
auto end = std::chrono::system_clock::now();
53+
auto end = std::chrono::system_clock::now();
5054
std::chrono::duration<double> elapsed_seconds = end - start;
5155

5256
std::cout << "Vamana index has " << index.size() << " vectors" << std::endl;
53-
std::cout << "Vamana graph has degree " << index.graph_degree()
54-
<< ", graph size [" << index.graph().extent(0) << ", "
55-
<< index.graph().extent(1) << "]" << std::endl;
57+
std::cout << "Vamana graph has degree " << index.graph_degree() << ", graph size ["
58+
<< index.graph().extent(0) << ", " << index.graph().extent(1) << "]" << std::endl;
5659

5760
std::cout << "Time to build index: " << elapsed_seconds.count() << "s\n";
5861

5962
// Output index to file
6063
serialize(dev_resources, out_fname, index);
6164
}
6265

63-
void usage() {
64-
printf("Usage: ./vamana_example <data filename> <output filename> <graph "
65-
"degree> <visited_size> <max_fraction> <iterations> \n");
66+
void usage()
67+
{
68+
printf(
69+
"Usage: ./vamana_example <data filename> <output filename> <graph "
70+
"degree> <visited_size> <max_fraction> <iterations> \n");
6671
printf("Input file expected to be binary file of fp32 vectors.\n");
6772
printf("Graph degree sizes supported: 32, 64, 128, 256\n");
6873
printf("Visited_size must be > degree and a power of 2.\n");
@@ -71,13 +76,14 @@ void usage() {
7176
exit(1);
7277
}
7378

74-
int main(int argc, char *argv[]) {
79+
int main(int argc, char* argv[])
80+
{
7581
raft::device_resources dev_resources;
7682

7783
// Set pool memory resource with 1 GiB initial pool size. All allocations use
7884
// the same pool.
7985
rmm::mr::pool_memory_resource<rmm::mr::device_memory_resource> pool_mr(
80-
rmm::mr::get_current_device_resource(), 1024 * 1024 * 1024ull);
86+
rmm::mr::get_current_device_resource(), 1024 * 1024 * 1024ull);
8187
rmm::mr::set_current_device_resource(&pool_mr);
8288

8389
// Alternatively, one could define a pool allocator for temporary arrays (used
@@ -87,22 +93,24 @@ int main(int argc, char *argv[]) {
8793
// limit. raft::resource::set_workspace_to_pool_resource(dev_resources, 2 *
8894
// 1024 * 1024 * 1024ull);
8995

90-
if (argc != 7)
91-
usage();
96+
if (argc != 7) usage();
9297

93-
std::string data_fname = (std::string)(argv[1]); // Input filename
94-
std::string out_fname = (std::string)argv[2]; // Output index filename
95-
int degree = atoi(argv[3]);
96-
int max_visited = atoi(argv[4]);
97-
float max_fraction = atof(argv[5]);
98-
int iters = atoi(argv[6]);
98+
std::string data_fname = (std::string)(argv[1]); // Input filename
99+
std::string out_fname = (std::string)argv[2]; // Output index filename
100+
int degree = atoi(argv[3]);
101+
int max_visited = atoi(argv[4]);
102+
float max_fraction = atof(argv[5]);
103+
int iters = atoi(argv[6]);
99104

100105
// Read in binary dataset file
101-
auto dataset =
102-
read_bin_dataset<uint8_t, int64_t>(dev_resources, data_fname, INT_MAX);
106+
auto dataset = read_bin_dataset<uint8_t, int64_t>(dev_resources, data_fname, INT_MAX);
103107

104108
// Simple build example to create graph and write to a file
105-
vamana_build_and_write<uint8_t>(
106-
dev_resources, raft::make_const_mdspan(dataset.view()), out_fname, degree,
107-
max_visited, max_fraction, iters);
109+
vamana_build_and_write<uint8_t>(dev_resources,
110+
raft::make_const_mdspan(dataset.view()),
111+
out_fname,
112+
degree,
113+
max_visited,
114+
max_fraction,
115+
iters);
108116
}

0 commit comments

Comments
 (0)