GENIE
match_integrated.cu File Reference
#include <algorithm>
#include <iomanip>
#include <iostream>
#include <stdlib.h>
#include <string>
#include <sstream>
#include <math.h>
#include <thrust/copy.h>
#include <thrust/device_vector.h>
#include <genie/utility/cuda_macros.h>
#include <genie/utility/Logger.h>
#include <genie/utility/PerfLogger.h>
#include <genie/utility/Timing.h>
#include <genie/exception/exception.h>
#include <genie/compression/DeviceCodecFactory.h>
#include <genie/compression/DeviceCompositeCodec.h>
#include <genie/compression/DeviceSerialCodec.h>
#include <genie/compression/DeviceBitPackingCodec.h>
#include <genie/compression/DeviceVarintCodec.h>
#include "match_common.h"
#include "match_integrated.h"
#include "match_device_utils.h"

Go to the source code of this file.

Namespaces

 genie
 This is the top-level namespace of the project.
 
 genie::matching
 

Macros

#define GPUGENIE_INTEGRATED_KERNEL_SM_SIZE   (1024)
 

Functions

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 > &)
 
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 > &)
 
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 > &)
 
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 > &)
 
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 > &)
 
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 > &)
 
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 > &)
 
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 > &)
 
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 > &)
 
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 > &)
 
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 > &)
 
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 > &)
 
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 > &)
 
int genie::matching::getBitmapSize (int &in_out_bitmap_bits, u32 in_shift_bits_subsequence, int in_number_of_data_points, int in_queries_size)
 
int genie::matching::build_compressed_queries (vector< Query > &queries, inv_compr_table *ctable, vector< Query::dim > &dims, int max_load)
 
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)
 

Macro Definition Documentation

◆ GPUGENIE_INTEGRATED_KERNEL_SM_SIZE

#define GPUGENIE_INTEGRATED_KERNEL_SM_SIZE   (1024)

Maximal length the codecs are able to decompress into.

GENIE uses fixed 256 threads in its kernels. This implies that a Codec has to have a thread load at least 4 (one thread decompressed into 4 values), otherwise such codec will fail.

Definition at line 35 of file match_integrated.cu.