GENIE
Classes | Functions | Variables
genie::matching Namespace Reference

Classes

struct  data_t
 

Functions

void heap_count_topk (thrust::device_vector< genie::matching::data_t > &d_data, thrust::device_vector< genie::matching::data_t > &d_topk, thrust::device_vector< u32 > &d_threshold, thrust::device_vector< u32 > &d_passCount, int topk, int num_of_queries)
 
SearchResult Match (const std::shared_ptr< const genie::table::inv_table > &table, const std::vector< genie::query::Query > &queries, const uint32_t dim, const uint32_t k)
 
void knn (genie::table::inv_table &table, std::vector< genie::query::Query > &queries, thrust::device_vector< int > &d_top_indexes, thrust::device_vector< int > &d_top_count, int hash_table_size, int max_load, int bitmap_bits)
 
void knn_MT (std::vector< genie::table::inv_table *> &table, std::vector< std::vector< genie::query::Query > > &queries, std::vector< thrust::device_vector< int > > &d_top_indexes, std::vector< thrust::device_vector< int > > &d_top_count, std::vector< int > &hash_table_size, std::vector< int > &max_load, int bitmap_bits)
 
void knn_bijectMap (genie::table::inv_table &table, std::vector< genie::query::Query > &queries, thrust::device_vector< int > &d_top_indexes, thrust::device_vector< int > &d_top_count, int hash_table_size, int max_load, int bitmap_bits)
 
void knn_bijectMap_MT (std::vector< genie::table::inv_table *> &table, std::vector< std::vector< genie::query::Query > > &queries, std::vector< thrust::device_vector< int > > &d_top_indexes, std::vector< thrust::device_vector< int > > &d_top_count, std::vector< int > &hash_table_size, std::vector< int > &max_load, int bitmap_bits)
 
__global__ void match_AT (int m_size, int i_size, int hash_table_size, int *d_inv, Query::dim *d_dims, T_HASHTABLE *hash_table_list, u32 *bitmap_list, int bitmap_bits, u32 *d_topks, u32 *d_threshold, u32 *d_passCount, u32 num_of_max_count, u32 *noiih, bool *overflow, unsigned int shift_bits_subsequence)
 
int build_queries (vector< Query > &queries, inv_table &table, vector< Query::dim > &dims, int max_load)
 
int cal_max_topk (vector< Query > &queries)
 
void match (inv_table &table, vector< Query > &queries, device_vector< data_t > &d_data, device_vector< u32 > &d_bitmap, int hash_table_size, int max_load, int bitmap_bits, device_vector< u32 > &d_noiih, device_vector< u32 > &d_threshold, device_vector< u32 > &d_passCount)
 
void match_MT (vector< inv_table *> &table, vector< vector< Query > > &queries, vector< device_vector< data_t > > &d_data, vector< device_vector< u32 > > &d_bitmap, vector< int > &hash_table_size, vector< int > &max_load, int bitmap_bits, vector< device_vector< u32 > > &d_noiih, vector< device_vector< u32 > > &d_threshold, vector< device_vector< u32 > > &d_passCount, size_t start, size_t finish)
 
int cal_max_topk (std::vector< genie::query::Query > &queries)
 
void match (genie::table::inv_table &table, std::vector< genie::query::Query > &queries, thrust::device_vector< genie::matching::data_t > &d_data, thrust::device_vector< u32 > &d_bitmap, int hash_table_size, int max_load, int bitmap_bits, thrust::device_vector< u32 > &d_noiih, thrust::device_vector< u32 > &d_threshold, thrust::device_vector< u32 > &d_passCount)
 Search the inv_table and save the match result into d_count and d_aggregation. More...
 
void match_MT (std::vector< genie::table::inv_table *> &table, std::vector< std::vector< genie::query::Query > > &queries, std::vector< thrust::device_vector< genie::matching::data_t > > &d_data, std::vector< thrust::device_vector< u32 > > &d_bitmap, std::vector< int > &hash_table_size, std::vector< int > &max_load, int bitmap_bits, std::vector< thrust::device_vector< u32 > > &d_noiih, std::vector< thrust::device_vector< u32 > > &d_threshold, std::vector< thrust::device_vector< u32 > > &d_passCount, size_t start, size_t finish)
 
int build_queries (std::vector< genie::query::Query > &queries, genie::table::inv_table &table, std::vector< genie::query::Query::dim > &dims, int max_load)
 
__global__ void convert_to_data (T_HASHTABLE *table, u32 size)
 
__forceinline__ __host__ __device__ T_KEY get_key_pos (T_HASHTABLE key)
 
__forceinline__ __host__ __device__ T_AGE get_key_age (T_HASHTABLE key)
 
__host__ __forceinline__ __device__ u32 get_key_attach_id (T_HASHTABLE key)
 
__host__ __forceinline__ __device__ T_HASHTABLE pack_key_pos (T_KEY p)
 
__host__ __forceinline__ __device__ T_HASHTABLE pack_key_pos_and_attach_id_and_age (T_KEY p, u32 i, T_AGE a)
 
__forceinline__ __device__ u32 hash (T_KEY key, T_AGE age, int hash_table_size)
 
__forceinline__ __device__ __host__ void print_binary (char *b, u32 data)
 
__forceinline__ __device__ __host__ u32 get_count (u32 data, int offset, int bits)
 
__forceinline__ __device__ __host__ u32 pack_count (u32 data, int offset, int bits, u32 count)
 
__forceinline__ __device__ void access_kernel (u32 id, T_HASHTABLE *htable, int hash_table_size, genie::query::Query::dim &q, bool *key_found)
 
__device__ __forceinline__ void access_kernel_AT (u32 id, T_HASHTABLE *htable, int hash_table_size, genie::query::Query::dim &q, u32 count, bool *key_found, u32 *my_threshold, bool *pass_threshold)
 
__device__ __forceinline__ void hash_kernel_AT (u32 id, T_HASHTABLE *htable, int hash_table_size, genie::query::Query::dim &q, u32 count, u32 *my_threshold, u32 *my_noiih, bool *overflow, bool *pass_threshold)
 
__device__ __forceinline__ u32 bitmap_kernel_AT (u32 access_id, u32 *bitmap, int bits, int my_threshold, bool *key_eligible)
 
__device__ __forceinline__ void updateThreshold (u32 *my_passCount, u32 *my_threshold, u32 my_topk, u32 count)
 
template void match_integrated< DeviceCopyCodec > (inv_compr_table &, std::vector< Query > &, thrust::device_vector< data_t > &, thrust::device_vector< u32 > &, int, int, thrust::device_vector< u32 > &, thrust::device_vector< u32 > &, thrust::device_vector< u32 > &)
 
template void match_integrated< DeviceDeltaCodec > (inv_compr_table &, std::vector< Query > &, thrust::device_vector< data_t > &, thrust::device_vector< u32 > &, int, int, thrust::device_vector< u32 > &, thrust::device_vector< u32 > &, thrust::device_vector< u32 > &)
 
template void match_integrated< DeviceBitPackingCodec > (inv_compr_table &, std::vector< Query > &, thrust::device_vector< data_t > &, thrust::device_vector< u32 > &, int, int, thrust::device_vector< u32 > &, thrust::device_vector< u32 > &, thrust::device_vector< u32 > &)
 
template void match_integrated< DeviceVarintCodec > (inv_compr_table &, std::vector< Query > &, thrust::device_vector< data_t > &, thrust::device_vector< u32 > &, int, int, thrust::device_vector< u32 > &, thrust::device_vector< u32 > &, thrust::device_vector< u32 > &)
 
template void match_integrated< DeviceCompositeCodec< DeviceBitPackingCodec, DeviceCopyCodec > > (inv_compr_table &, std::vector< Query > &, thrust::device_vector< data_t > &, thrust::device_vector< u32 > &, int, int, thrust::device_vector< u32 > &, thrust::device_vector< u32 > &, thrust::device_vector< u32 > &)
 
template void match_integrated< DeviceCompositeCodec< DeviceBitPackingCodec, DeviceVarintCodec > > (inv_compr_table &, std::vector< Query > &, thrust::device_vector< data_t > &, thrust::device_vector< u32 > &, int, int, thrust::device_vector< u32 > &, thrust::device_vector< u32 > &, thrust::device_vector< u32 > &)
 
template void match_integrated< DeviceSerialCodec< DeviceCopyCodec, DeviceCopyCodec > > (inv_compr_table &, std::vector< Query > &, thrust::device_vector< data_t > &, thrust::device_vector< u32 > &, int, int, thrust::device_vector< u32 > &, thrust::device_vector< u32 > &, thrust::device_vector< u32 > &)
 
template void match_integrated< DeviceSerialCodec< DeviceDeltaCodec, DeviceCopyCodec > > (inv_compr_table &, std::vector< Query > &, thrust::device_vector< data_t > &, thrust::device_vector< u32 > &, int, int, thrust::device_vector< u32 > &, thrust::device_vector< u32 > &, thrust::device_vector< u32 > &)
 
template void match_integrated< DeviceSerialCodec< DeviceDeltaCodec, DeviceDeltaCodec > > (inv_compr_table &, std::vector< Query > &, thrust::device_vector< data_t > &, thrust::device_vector< u32 > &, int, int, thrust::device_vector< u32 > &, thrust::device_vector< u32 > &, thrust::device_vector< u32 > &)
 
template void match_integrated< DeviceSerialCodec< DeviceDeltaCodec, DeviceVarintCodec > > (inv_compr_table &, std::vector< Query > &, thrust::device_vector< data_t > &, thrust::device_vector< u32 > &, int, int, thrust::device_vector< u32 > &, thrust::device_vector< u32 > &, thrust::device_vector< u32 > &)
 
template void match_integrated< DeviceSerialCodec< DeviceDeltaCodec, DeviceBitPackingCodec > > (inv_compr_table &, std::vector< Query > &, thrust::device_vector< data_t > &, thrust::device_vector< u32 > &, int, int, thrust::device_vector< u32 > &, thrust::device_vector< u32 > &, thrust::device_vector< u32 > &)
 
template void match_integrated< DeviceSerialCodec< DeviceDeltaCodec, DeviceCompositeCodec< DeviceBitPackingCodec, DeviceCopyCodec > > > (inv_compr_table &, std::vector< Query > &, thrust::device_vector< data_t > &, thrust::device_vector< u32 > &, int, int, thrust::device_vector< u32 > &, thrust::device_vector< u32 > &, thrust::device_vector< u32 > &)
 
template void match_integrated< DeviceSerialCodec< DeviceDeltaCodec, DeviceCompositeCodec< DeviceBitPackingCodec, DeviceVarintCodec > > > (inv_compr_table &, std::vector< Query > &, thrust::device_vector< data_t > &, thrust::device_vector< u32 > &, int, int, thrust::device_vector< u32 > &, thrust::device_vector< u32 > &, thrust::device_vector< u32 > &)
 
int getBitmapSize (int &in_out_bitmap_bits, u32 in_shift_bits_subsequence, int in_number_of_data_points, int in_queries_size)
 
int build_compressed_queries (vector< Query > &queries, inv_compr_table *ctable, vector< Query::dim > &dims, int max_load)
 
template<class Codec >
__global__ void match_adaptiveThreshold_integrated (int m_size, int i_size, int hash_table_size, uint32_t *d_compr_inv, Query::dim *d_dims, T_HASHTABLE *hash_table_list, u32 *bitmap_list, int bitmap_bits, u32 *d_topks, u32 *d_threshold, u32 *d_passCount, u32 num_of_max_count, u32 *noiih, bool *overflow, unsigned int shift_bits_subsequence)
 
template<class Codec >
void match_integrated (genie::table::inv_compr_table &table, std::vector< genie::query::Query > &queries, thrust::device_vector< genie::matching::data_t > &d_data, thrust::device_vector< u32 > &d_bitmap, int hash_table_size, int bitmap_bits, thrust::device_vector< u32 > &d_noiih, thrust::device_vector< u32 > &d_threshold, thrust::device_vector< u32 > &d_passCount)
 

Variables

const size_t MATCH_THREADS_PER_BLOCK = 256
 
const T_AGE MAX_AGE = 16u
 
const uint32_t KEY_TYPE_BITS = 28u
 
const uint32_t KEY_TYPE_MASK = u32(u64((1ull) << KEY_TYPE_BITS) - 1u)
 
const uint32_t ATTACH_ID_TYPE_BITS = 32u
 
const uint32_t ATTACH_ID_TYPE_MASK = u32(u64((1ull) << ATTACH_ID_TYPE_BITS) - 1ul)
 
const uint32_t KEY_TYPE_INIT_AGE = 1u
 
const uint32_t KEY_TYPE_NULL_AGE = 0u
 

Function Documentation

◆ access_kernel()

__forceinline__ __device__ void genie::matching::access_kernel ( u32  id,
T_HASHTABLE *  htable,
int  hash_table_size,
genie::query::Query::dim &  q,
bool *  key_found 
)

Definition at line 94 of file match_device_utils.h.

◆ access_kernel_AT()

__device__ __forceinline__ void genie::matching::access_kernel_AT ( u32  id,
T_HASHTABLE *  htable,
int  hash_table_size,
genie::query::Query::dim &  q,
u32  count,
bool *  key_found,
u32 *  my_threshold,
bool *  pass_threshold 
)

Definition at line 161 of file match_device_utils.h.

◆ bitmap_kernel_AT()

__device__ __forceinline__ u32 genie::matching::bitmap_kernel_AT ( u32  access_id,
u32 *  bitmap,
int  bits,
int  my_threshold,
bool *  key_eligible 
)

Definition at line 423 of file match_device_utils.h.

◆ build_compressed_queries()

int genie::matching::build_compressed_queries ( vector< Query > &  queries,
inv_compr_table *  ctable,
vector< Query::dim > &  dims,
int  max_load 
)

Definition at line 142 of file match_integrated.cu.

◆ build_queries() [1/2]

int genie::matching::build_queries ( std::vector< genie::query::Query > &  queries,
genie::table::inv_table &  table,
std::vector< genie::query::Query::dim > &  dims,
int  max_load 
)

◆ build_queries() [2/2]

int genie::matching::build_queries ( vector< Query > &  queries,
inv_table &  table,
vector< Query::dim > &  dims,
int  max_load 
)

Definition at line 161 of file match.cu.

◆ cal_max_topk() [1/2]

int genie::matching::cal_max_topk ( std::vector< genie::query::Query > &  queries)

◆ cal_max_topk() [2/2]

int genie::matching::cal_max_topk ( vector< Query > &  queries)

Definition at line 212 of file match.cu.

◆ convert_to_data()

__global__ void genie::matching::convert_to_data ( T_HASHTABLE *  table,
u32  size 
)

Definition at line 10 of file match_common.cu.

◆ get_count()

__forceinline__ __device__ __host__ u32 genie::matching::get_count ( u32  data,
int  offset,
int  bits 
)

Definition at line 80 of file match_device_utils.h.

◆ get_key_age()

__forceinline__ __host__ __device__ T_AGE genie::matching::get_key_age ( T_HASHTABLE  key)

Definition at line 41 of file match_device_utils.h.

◆ get_key_attach_id()

__host__ __forceinline__ __device__ u32 genie::matching::get_key_attach_id ( T_HASHTABLE  key)

Definition at line 47 of file match_device_utils.h.

◆ get_key_pos()

__forceinline__ __host__ __device__ T_KEY genie::matching::get_key_pos ( T_HASHTABLE  key)

Definition at line 36 of file match_device_utils.h.

◆ getBitmapSize()

int genie::matching::getBitmapSize ( int &  in_out_bitmap_bits,
u32  in_shift_bits_subsequence,
int  in_number_of_data_points,
int  in_queries_size 
)

Definition at line 106 of file match_integrated.cu.

◆ hash()

__forceinline__ __device__ u32 genie::matching::hash ( T_KEY  key,
T_AGE  age,
int  hash_table_size 
)

Definition at line 65 of file match_device_utils.h.

◆ hash_kernel_AT()

__device__ __forceinline__ void genie::matching::hash_kernel_AT ( u32  id,
T_HASHTABLE *  htable,
int  hash_table_size,
genie::query::Query::dim &  q,
u32  count,
u32 *  my_threshold,
u32 *  my_noiih,
bool *  overflow,
bool *  pass_threshold 
)

Definition at line 260 of file match_device_utils.h.

◆ heap_count_topk()

void genie::matching::heap_count_topk ( thrust::device_vector< genie::matching::data_t > &  d_data,
thrust::device_vector< genie::matching::data_t > &  d_topk,
thrust::device_vector< u32 > &  d_threshold,
thrust::device_vector< u32 > &  d_passCount,
int  topk,
int  num_of_queries 
)

◆ knn()

void genie::matching::knn ( genie::table::inv_table &  table,
std::vector< genie::query::Query > &  queries,
thrust::device_vector< int > &  d_top_indexes,
thrust::device_vector< int > &  d_top_count,
int  hash_table_size,
int  max_load,
int  bitmap_bits 
)

◆ knn_bijectMap()

void genie::matching::knn_bijectMap ( genie::table::inv_table &  table,
std::vector< genie::query::Query > &  queries,
thrust::device_vector< int > &  d_top_indexes,
thrust::device_vector< int > &  d_top_count,
int  hash_table_size,
int  max_load,
int  bitmap_bits 
)

◆ knn_bijectMap_MT()

void genie::matching::knn_bijectMap_MT ( std::vector< genie::table::inv_table *> &  table,
std::vector< std::vector< genie::query::Query > > &  queries,
std::vector< thrust::device_vector< int > > &  d_top_indexes,
std::vector< thrust::device_vector< int > > &  d_top_count,
std::vector< int > &  hash_table_size,
std::vector< int > &  max_load,
int  bitmap_bits 
)

◆ knn_MT()

void genie::matching::knn_MT ( std::vector< genie::table::inv_table *> &  table,
std::vector< std::vector< genie::query::Query > > &  queries,
std::vector< thrust::device_vector< int > > &  d_top_indexes,
std::vector< thrust::device_vector< int > > &  d_top_count,
std::vector< int > &  hash_table_size,
std::vector< int > &  max_load,
int  bitmap_bits 
)

◆ Match()

SearchResult genie::matching::Match ( const std::shared_ptr< const genie::table::inv_table > &  table,
const std::vector< genie::query::Query > &  queries,
const uint32_t  dim,
const uint32_t  k 
)

◆ match() [1/2]

void genie::matching::match ( genie::table::inv_table &  table,
std::vector< genie::query::Query > &  queries,
thrust::device_vector< genie::matching::data_t > &  d_data,
thrust::device_vector< u32 > &  d_bitmap,
int  hash_table_size,
int  max_load,
int  bitmap_bits,
thrust::device_vector< u32 > &  d_noiih,
thrust::device_vector< u32 > &  d_threshold,
thrust::device_vector< u32 > &  d_passCount 
)

Search the inv_table and save the match result into d_count and d_aggregation.

Parameters
tableThe inv_table which will be searched.
queriesThe queries.
d_dataThe output data consisting of count and the index of the data in table.
hash_table_sizeThe hash table size.
max_loadThe maximum number of posting list items that can be processed by one gpu block
bitmap_bitsThe threshold for the count heap
d_noiihThe number of items in hash table
d_thresholdThe container for heap-count thresholds of each query.
d_passCountThe container for heap-count counts in each buckets of each query.

◆ match() [2/2]

void genie::matching::match ( inv_table &  table,
vector< Query > &  queries,
device_vector< data_t > &  d_data,
device_vector< u32 > &  d_bitmap,
int  hash_table_size,
int  max_load,
int  bitmap_bits,
device_vector< u32 > &  d_noiih,
device_vector< u32 > &  d_threshold,
device_vector< u32 > &  d_passCount 
)

Definition at line 224 of file match.cu.

◆ match_adaptiveThreshold_integrated()

template<class Codec >
__global__ void genie::matching::match_adaptiveThreshold_integrated ( int  m_size,
int  i_size,
int  hash_table_size,
uint32_t *  d_compr_inv,
Query::dim *  d_dims,
T_HASHTABLE *  hash_table_list,
u32 *  bitmap_list,
int  bitmap_bits,
u32 *  d_topks,
u32 *  d_threshold,
u32 *  d_passCount,
u32  num_of_max_count,
u32 *  noiih,
bool *  overflow,
unsigned int  shift_bits_subsequence 
)

Definition at line 169 of file match_integrated.cu.

◆ match_AT()

__global__ void genie::matching::match_AT ( int  m_size,
int  i_size,
int  hash_table_size,
int *  d_inv,
Query::dim *  d_dims,
T_HASHTABLE *  hash_table_list,
u32 *  bitmap_list,
int  bitmap_bits,
u32 *  d_topks,
u32 *  d_threshold,
u32 *  d_passCount,
u32  num_of_max_count,
u32 *  noiih,
bool *  overflow,
unsigned int  shift_bits_subsequence 
)

Definition at line 44 of file match.cu.

◆ match_integrated()

template<class Codec >
void genie::matching::match_integrated ( genie::table::inv_compr_table &  table,
std::vector< genie::query::Query > &  queries,
thrust::device_vector< genie::matching::data_t > &  d_data,
thrust::device_vector< u32 > &  d_bitmap,
int  hash_table_size,
int  bitmap_bits,
thrust::device_vector< u32 > &  d_noiih,
thrust::device_vector< u32 > &  d_threshold,
thrust::device_vector< u32 > &  d_passCount 
)

◆ match_integrated< DeviceBitPackingCodec >()

template void genie::matching::match_integrated< DeviceBitPackingCodec > ( inv_compr_table &  ,
std::vector< Query > &  ,
thrust::device_vector< data_t > &  ,
thrust::device_vector< u32 > &  ,
int  ,
int  ,
thrust::device_vector< u32 > &  ,
thrust::device_vector< u32 > &  ,
thrust::device_vector< u32 > &   
)

◆ match_integrated< DeviceCompositeCodec< DeviceBitPackingCodec, DeviceCopyCodec > >()

template void genie::matching::match_integrated< DeviceCompositeCodec< DeviceBitPackingCodec, DeviceCopyCodec > > ( inv_compr_table &  ,
std::vector< Query > &  ,
thrust::device_vector< data_t > &  ,
thrust::device_vector< u32 > &  ,
int  ,
int  ,
thrust::device_vector< u32 > &  ,
thrust::device_vector< u32 > &  ,
thrust::device_vector< u32 > &   
)

◆ match_integrated< DeviceCompositeCodec< DeviceBitPackingCodec, DeviceVarintCodec > >()

template void genie::matching::match_integrated< DeviceCompositeCodec< DeviceBitPackingCodec, DeviceVarintCodec > > ( inv_compr_table &  ,
std::vector< Query > &  ,
thrust::device_vector< data_t > &  ,
thrust::device_vector< u32 > &  ,
int  ,
int  ,
thrust::device_vector< u32 > &  ,
thrust::device_vector< u32 > &  ,
thrust::device_vector< u32 > &   
)

◆ match_integrated< DeviceCopyCodec >()

template void genie::matching::match_integrated< DeviceCopyCodec > ( inv_compr_table &  ,
std::vector< Query > &  ,
thrust::device_vector< data_t > &  ,
thrust::device_vector< u32 > &  ,
int  ,
int  ,
thrust::device_vector< u32 > &  ,
thrust::device_vector< u32 > &  ,
thrust::device_vector< u32 > &   
)

◆ match_integrated< DeviceDeltaCodec >()

template void genie::matching::match_integrated< DeviceDeltaCodec > ( inv_compr_table &  ,
std::vector< Query > &  ,
thrust::device_vector< data_t > &  ,
thrust::device_vector< u32 > &  ,
int  ,
int  ,
thrust::device_vector< u32 > &  ,
thrust::device_vector< u32 > &  ,
thrust::device_vector< u32 > &   
)

◆ match_integrated< DeviceSerialCodec< DeviceCopyCodec, DeviceCopyCodec > >()

template void genie::matching::match_integrated< DeviceSerialCodec< DeviceCopyCodec, DeviceCopyCodec > > ( inv_compr_table &  ,
std::vector< Query > &  ,
thrust::device_vector< data_t > &  ,
thrust::device_vector< u32 > &  ,
int  ,
int  ,
thrust::device_vector< u32 > &  ,
thrust::device_vector< u32 > &  ,
thrust::device_vector< u32 > &   
)

◆ match_integrated< DeviceSerialCodec< DeviceDeltaCodec, DeviceBitPackingCodec > >()

template void genie::matching::match_integrated< DeviceSerialCodec< DeviceDeltaCodec, DeviceBitPackingCodec > > ( inv_compr_table &  ,
std::vector< Query > &  ,
thrust::device_vector< data_t > &  ,
thrust::device_vector< u32 > &  ,
int  ,
int  ,
thrust::device_vector< u32 > &  ,
thrust::device_vector< u32 > &  ,
thrust::device_vector< u32 > &   
)

◆ match_integrated< DeviceSerialCodec< DeviceDeltaCodec, DeviceCompositeCodec< DeviceBitPackingCodec, DeviceCopyCodec > > >()

template void genie::matching::match_integrated< DeviceSerialCodec< DeviceDeltaCodec, DeviceCompositeCodec< DeviceBitPackingCodec, DeviceCopyCodec > > > ( inv_compr_table &  ,
std::vector< Query > &  ,
thrust::device_vector< data_t > &  ,
thrust::device_vector< u32 > &  ,
int  ,
int  ,
thrust::device_vector< u32 > &  ,
thrust::device_vector< u32 > &  ,
thrust::device_vector< u32 > &   
)

◆ match_integrated< DeviceSerialCodec< DeviceDeltaCodec, DeviceCompositeCodec< DeviceBitPackingCodec, DeviceVarintCodec > > >()

template void genie::matching::match_integrated< DeviceSerialCodec< DeviceDeltaCodec, DeviceCompositeCodec< DeviceBitPackingCodec, DeviceVarintCodec > > > ( inv_compr_table &  ,
std::vector< Query > &  ,
thrust::device_vector< data_t > &  ,
thrust::device_vector< u32 > &  ,
int  ,
int  ,
thrust::device_vector< u32 > &  ,
thrust::device_vector< u32 > &  ,
thrust::device_vector< u32 > &   
)

◆ match_integrated< DeviceSerialCodec< DeviceDeltaCodec, DeviceCopyCodec > >()

template void genie::matching::match_integrated< DeviceSerialCodec< DeviceDeltaCodec, DeviceCopyCodec > > ( inv_compr_table &  ,
std::vector< Query > &  ,
thrust::device_vector< data_t > &  ,
thrust::device_vector< u32 > &  ,
int  ,
int  ,
thrust::device_vector< u32 > &  ,
thrust::device_vector< u32 > &  ,
thrust::device_vector< u32 > &   
)

◆ match_integrated< DeviceSerialCodec< DeviceDeltaCodec, DeviceDeltaCodec > >()

template void genie::matching::match_integrated< DeviceSerialCodec< DeviceDeltaCodec, DeviceDeltaCodec > > ( inv_compr_table &  ,
std::vector< Query > &  ,
thrust::device_vector< data_t > &  ,
thrust::device_vector< u32 > &  ,
int  ,
int  ,
thrust::device_vector< u32 > &  ,
thrust::device_vector< u32 > &  ,
thrust::device_vector< u32 > &   
)

◆ match_integrated< DeviceSerialCodec< DeviceDeltaCodec, DeviceVarintCodec > >()

template void genie::matching::match_integrated< DeviceSerialCodec< DeviceDeltaCodec, DeviceVarintCodec > > ( inv_compr_table &  ,
std::vector< Query > &  ,
thrust::device_vector< data_t > &  ,
thrust::device_vector< u32 > &  ,
int  ,
int  ,
thrust::device_vector< u32 > &  ,
thrust::device_vector< u32 > &  ,
thrust::device_vector< u32 > &   
)

◆ match_integrated< DeviceVarintCodec >()

template void genie::matching::match_integrated< DeviceVarintCodec > ( inv_compr_table &  ,
std::vector< Query > &  ,
thrust::device_vector< data_t > &  ,
thrust::device_vector< u32 > &  ,
int  ,
int  ,
thrust::device_vector< u32 > &  ,
thrust::device_vector< u32 > &  ,
thrust::device_vector< u32 > &   
)

◆ match_MT() [1/2]

void genie::matching::match_MT ( std::vector< genie::table::inv_table *> &  table,
std::vector< std::vector< genie::query::Query > > &  queries,
std::vector< thrust::device_vector< genie::matching::data_t > > &  d_data,
std::vector< thrust::device_vector< u32 > > &  d_bitmap,
std::vector< int > &  hash_table_size,
std::vector< int > &  max_load,
int  bitmap_bits,
std::vector< thrust::device_vector< u32 > > &  d_noiih,
std::vector< thrust::device_vector< u32 > > &  d_threshold,
std::vector< thrust::device_vector< u32 > > &  d_passCount,
size_t  start,
size_t  finish 
)

◆ match_MT() [2/2]

void genie::matching::match_MT ( vector< inv_table *> &  table,
vector< vector< Query > > &  queries,
vector< device_vector< data_t > > &  d_data,
vector< device_vector< u32 > > &  d_bitmap,
vector< int > &  hash_table_size,
vector< int > &  max_load,
int  bitmap_bits,
vector< device_vector< u32 > > &  d_noiih,
vector< device_vector< u32 > > &  d_threshold,
vector< device_vector< u32 > > &  d_passCount,
size_t  start,
size_t  finish 
)

Definition at line 502 of file match.cu.

◆ pack_count()

__forceinline__ __device__ __host__ u32 genie::matching::pack_count ( u32  data,
int  offset,
int  bits,
u32  count 
)

Definition at line 86 of file match_device_utils.h.

◆ pack_key_pos()

__host__ __forceinline__ __device__ T_HASHTABLE genie::matching::pack_key_pos ( T_KEY  p)

Definition at line 52 of file match_device_utils.h.

◆ pack_key_pos_and_attach_id_and_age()

__host__ __forceinline__ __device__ T_HASHTABLE genie::matching::pack_key_pos_and_attach_id_and_age ( T_KEY  p,
u32  i,
T_AGE  a 
)

Definition at line 57 of file match_device_utils.h.

◆ print_binary()

__forceinline__ __device__ __host__ void genie::matching::print_binary ( char *  b,
u32  data 
)

Definition at line 72 of file match_device_utils.h.

◆ updateThreshold()

__device__ __forceinline__ void genie::matching::updateThreshold ( u32 *  my_passCount,
u32 *  my_threshold,
u32  my_topk,
u32  count 
)

Definition at line 449 of file match_device_utils.h.

Variable Documentation

◆ ATTACH_ID_TYPE_BITS

const uint32_t genie::matching::ATTACH_ID_TYPE_BITS = 32u

Definition at line 25 of file match_device_utils.h.

◆ ATTACH_ID_TYPE_MASK

const uint32_t genie::matching::ATTACH_ID_TYPE_MASK = u32(u64((1ull) << ATTACH_ID_TYPE_BITS) - 1ul)

Definition at line 26 of file match_device_utils.h.

◆ KEY_TYPE_BITS

const uint32_t genie::matching::KEY_TYPE_BITS = 28u

Definition at line 23 of file match_device_utils.h.

◆ KEY_TYPE_INIT_AGE

const uint32_t genie::matching::KEY_TYPE_INIT_AGE = 1u

Definition at line 27 of file match_device_utils.h.

◆ KEY_TYPE_MASK

const uint32_t genie::matching::KEY_TYPE_MASK = u32(u64((1ull) << KEY_TYPE_BITS) - 1u)

Definition at line 24 of file match_device_utils.h.

◆ KEY_TYPE_NULL_AGE

const uint32_t genie::matching::KEY_TYPE_NULL_AGE = 0u

Definition at line 28 of file match_device_utils.h.

◆ MATCH_THREADS_PER_BLOCK

const size_t genie::matching::MATCH_THREADS_PER_BLOCK = 256

Definition at line 41 of file match_common.h.

◆ MAX_AGE

const T_AGE genie::matching::MAX_AGE = 16u

Definition at line 22 of file match_device_utils.h.