2 #ifndef CAFFE2_UTILS_MIXED_UTILS_H 3 #define CAFFE2_UTILS_MIXED_UTILS_H 5 #include "caffe2/core/common_gpu.h" 6 #include "caffe2/core/context_gpu.h" 14 template <
typename T,
typename T2>
15 inline __device__ T mixed_mult(T data1, T2 data2) {
19 template <
typename T,
typename T2>
20 inline __device__ T mixed_add(T data1, T2 data2) {
24 template <
typename TIN,
typename TOUT>
25 inline __device__
void mixed_store(TIN* data_in, TOUT* data_out) {
31 inline __device__
void mixed_store(T* data_in, T* data_out) {
36 #ifdef CAFFE_HAS_CUDA_FP16 39 inline __device__
float mixed_mult(
float data1,
const float data2) {
44 inline __device__
float mixed_mult(
float data1,
const half data2) {
45 return data1 * __half2float(data2);
49 inline __device__
float mixed_mult(
float data1, float16 data2) {
50 half* data2_half =
reinterpret_cast<half*
>(&data2);
51 return data1 * __half2float(*data2_half);
54 inline __device__
float mixed_add(
float data1,
const float data2) {
59 inline __device__
float mixed_add(
float data1,
const half data2) {
60 return data1 + __half2float(data2);
64 inline __device__
float mixed_add(
float data1, float16 data2) {
65 half* data2_half =
reinterpret_cast<half*
>(&data2);
66 return data1 + __half2float(*data2_half);
70 inline __device__
void mixed_store(
float* data_in,
float* data_out) {
76 inline __device__
void mixed_store(half* data_in,
float* data_out) {
77 *data_out = __half2float(*data_in);
82 inline __device__
void mixed_store(float16* data_in,
float* data_out) {
83 half* data_in_half =
reinterpret_cast<half*
>(data_in);
84 *data_out = __half2float(*data_in_half);
89 inline __device__
void mixed_store(
float* data_in, float16* data_out) {
90 half data_in_half = __float2half(*data_in);
91 float16* data_in_float16 =
reinterpret_cast<float16*
>(&data_in_half);
92 *data_out = *data_in_float16;
97 inline __device__
void mixed_store(
float* data_in, half* data_out) {
98 half data_in_half = __float2half(*data_in);
99 *data_out = data_in_half;
102 #endif // for CAFFE_HAS_CUDA_FP16 104 #endif // for CAFFE2_UTILS_MIXED_UTILS_H A global dictionary that holds information about what Caffe2 modules have been loaded in the current ...