#include "env.h" #include "helper_cuda.h" #include <device_functions.h> #include "combo_device_tools.h"
typedef enum COMBO_DT_TYPE{
COMBO_DT_8C,
COMBO_DT_8U,
COMBO_DT_16S,
COMBO_DT_16U,
COMBO_DT_32S,
COMBO_DT_32U,
COMBO_DT_32F,
COMBO_DT_64F
}COMBO_DT_TYPE;
template<class T> static __device__ void print_data_device_t(uint8_t *data, int w, int h, int stride,int x, int y, int roi_w, int roi_h,int high_accuracy) { T* dataptr = (T*)data; for (int j = 0; j != roi_h; ++j) { for (int i = 0; i != roi_w; ++i) { int of = i+x + (j+y)*stride; if (high_accuracy) { printf("%.7f ", (float)dataptr[of]); } else { printf("%.2f ", (float)dataptr[of]); } } printf(" "); } printf(" "); } template<class T> static void print_data_host_t(uint8_t *data, int w, int h, int stride, int x, int y, int roi_w, int roi_h, int high_accuracy) { T* dataptr = (T*)data; for (int j = 0; j != roi_h; ++j) { for (int i = 0; i != roi_w; ++i) { int of = i + x + (j + y)*stride; if (high_accuracy) { printf("%.7f ", (float)(dataptr[of])); } else { printf("%.2f ", (float)(dataptr[of])); } } printf(" "); } printf(" "); } __global__ static void print_data_device(uint8_t *data, int w, int h, int stride,int x, int y, int roi_w, int roi_h,int high_accuracy, COMBO_DT_TYPE type) { void(*func)(uint8_t *data, int w, int h, int stride,int x, int y, int roi_w, int roi_h,int high_accuracy ); switch( type ){ case COMBO_DT_8C: func = print_data_device_t<int8_t>;break; case COMBO_DT_8U: func = print_data_device_t<uint8_t>;break; case COMBO_DT_16S: func = print_data_device_t<int16_t>;break; case COMBO_DT_16U: func = print_data_device_t<uint16_t>;break; case COMBO_DT_32S: func = print_data_device_t<int32_t>;break; case COMBO_DT_32U: func = print_data_device_t<uint32_t>;break; case COMBO_DT_32F: func = print_data_device_t<float>;break; case COMBO_DT_64F: func = print_data_device_t<double>;break; default: func = print_data_device_t<int8_t>;break; } func(data, w, h, stride, x, y, roi_w, roi_h, high_accuracy); } BEGIN_EXTERNAL_C void print_data_device(uint8_t *data,int w, int h, int stride, const char * str, int x, int y, int roi_w, int roi_h, int high_accuracy, COMBO_DT_TYPE type) { cudaThreadSynchronize(); if (!data) return; printf("%s :%p ", str, data); print_data_device << <1, 1 >> >(data, w, h, stride, x, y, roi_w, roi_h, high_accuracy, type); cudaThreadSynchronize(); } void print_data_host(uint8_t *data, int w, int h, int stride, const char * str, int x, int y, int roi_w, int roi_h, int high_accuracy, COMBO_DT_TYPE type) { if (!data) return; printf("%s :%p ", str, data); void(*func)(uint8_t *data, int w, int h, int stride, int x, int y, int roi_w, int roi_h, int high_accuracy); switch( type ){ case COMBO_DT_8C: func = print_data_host_t<int8_t>;break; case COMBO_DT_8U: func = print_data_host_t<uint8_t>;break; case COMBO_DT_16S: func = print_data_host_t<int16_t>;break; case COMBO_DT_16U: func = print_data_host_t<uint16_t>;break; case COMBO_DT_32S: func = print_data_host_t<int32_t>;break; case COMBO_DT_32U: func = print_data_host_t<uint32_t>;break; case COMBO_DT_32F: func = print_data_host_t<float>;break; case COMBO_DT_64F: func = print_data_host_t<double>;break; default: func = print_data_host_t<int8_t>;break; } func(data, w, h, stride, x, y, roi_w, roi_h, high_accuracy); } void print_mat_device(const float * data, int w, int h, int stride, const char * str, int x, int y, int roi_w, int roi_h, int high_accuracy) { print_data_device((uint8_t*)data, w, h, stride, str, x, y, roi_w, roi_h, high_accuracy, COMBO_DT_32F); } END_EXTERNAL_C