GPUMLib  0.2.2
GPU Machine Learning Library
CudaDefinitions.h
1 /*
2  Noel Lopes is a Professor at the Polytechnic of Guarda, Portugal
3  Copyright (C) 2009, 2010, 2011, 2012, 2013, 2014 Noel de Jesus Mendonça Lopes
4 
5  This file is part of GPUMLib.
6 
7  GPUMLib is free software: you can redistribute it and/or modify
8  it under the terms of the GNU General Public License as published by
9  the Free Software Foundation, either version 3 of the License, or
10  (at your option) any later version.
11 
12  This program is distributed in the hope that it will be useful,
13  but WITHOUT ANY WARRANTY; without even the implied warranty of
14  MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. See the
15  GNU General Public License for more details.
16 
17  You should have received a copy of the GNU General Public License
18  along with this program. If not, see <http://www.gnu.org/licenses/>.
19 */
20 
23 
24 #ifndef GPUMLib_CudaDefinitions_h
25 #define GPUMLib_CudaDefinitions_h
26 
27 #include <float.h>
28 #include <math.h>
29 
32 //#define FERMI
33 
36 #define USE_SINGLE_PRECISION_VARIABLES
37 
40 #define USE_SINGLE_PRECISION_FUNCTIONS
41 
43 #define KERNEL __global__ void
44 
45 #ifdef FERMI
46  #define MAX_THREADS_PER_BLOCK (1024)
48 #else
49  #define MAX_THREADS_PER_BLOCK (512)
51 #endif
52 
56 #define SIZE_SMALL_CUDA_VECTOR (3 * MAX_THREADS_PER_BLOCK)
57 
60 #define OPTIMAL_BLOCK_SIZE_REDUCTION (128)
61 
62 #ifdef USE_SINGLE_PRECISION_VARIABLES
63  typedef float cudafloat;
66 
70  #define CUDA_VALUE(X) (X##f)
71 
75  #define MAX_CUDAFLOAT (FLT_MAX)
76 
81  #define MIN_POSITIVE_CUDAFLOAT (FLT_MIN)
82 
88  #define MIN_CUDAFLOAT (-FLT_MAX)
89 #else // double precision variables and values
90  typedef double cudafloat;
93 
97  #define CUDA_VALUE(X) (X)
98 
102  #define MAX_CUDAFLOAT (DBL_MAX)
103 
108  #define MIN_POSITIVE_CUDAFLOAT (DBL_MIN)
109 
113  #define MIN_CUDAFLOAT (-DBL_MAX)
114 #endif
115 
116 #ifdef USE_SINGLE_PRECISION_FUNCTIONS
117  #define CUDA_EXP expf
120 
123  #define CUDA_SQRT sqrtf
124 
127  #define CUDA_TANH tanhf
128 
131  #define CUDA_COSH coshf
132 
135  #define CUDA_POW powf
136 
137 #else // double precision functions
138  #define CUDA_EXP exp
141 
144  #define CUDA_SQRT sqrt
145 
148  #define CUDA_TANH tanh
149 
152  #define CUDA_COSH cosh
153 
156  #define CUDA_POW pow
157 
158 #endif
159 
162 #define CUDA_SIGMOID(X) (CUDA_VALUE(1.0) / (CUDA_VALUE(1.0) + CUDA_EXP(-(X))))
163 
166 #define CUDA_SIGMOID_DERIVATE(OUTPUT) ((OUTPUT) * (CUDA_VALUE(1.0) - (OUTPUT)))
167 
169 #define SAME_DIRECTION(X, Y) (((X) > CUDA_VALUE(0.0) && (Y) > CUDA_VALUE(0.0)) || ((X) < CUDA_VALUE(0.0) && (Y) < CUDA_VALUE(0.0)))
170 
171 namespace GPUMLib {
172  #if defined(__CUDA_ARCH__)
173 
174  __device__ __forceinline__ bool IsInfOrNaN(cudafloat x) {
175  return (isnan(x) || isinf(x));
176  }
177 
178  __device__ __forceinline__ cudafloat Log(cudafloat x) {
179  if (x != 0) {
180  cudafloat y = log(x);
181  if (!IsInfOrNaN(y)) return y;
182  }
183 
184  return CUDA_VALUE(-7.0); //log(0.0000001)
185  }
186 
187  #else
188 
189  inline bool IsInfOrNaN(cudafloat x) {
190  #if (defined(_MSC_VER))
191  return (!_finite(x));
192  #else
193  return (isnan(x) || isinf(x));
194  #endif
195  }
196 
197  inline cudafloat Log(cudafloat x) {
198  if (x != 0) {
199  cudafloat y = log(x);
200  if (!IsInfOrNaN(y)) return y;
201  }
202 
203  return CUDA_VALUE(-7.0); //log(0.0000001)
204  }
205 
206  inline cudafloat AbsDiff(cudafloat a, cudafloat b) {
207  cudafloat d = a - b;
208  if (d < cudafloat(0.0)) return -d;
209  return d;
210  }
211 
212  #endif
213 }
214 
215 #endif
216 
#define CUDA_VALUE(X)
float cudafloat