Teuchos Package Browser (Single Doxygen Collection) Version of the Day
Teuchos_ScalarTraitsCUDA.hpp
Go to the documentation of this file.
00001 // @HEADER
00002 // ***********************************************************************
00003 //
00004 //                    Teuchos: Common Tools Package
00005 //                 Copyright (2004) Sandia Corporation
00006 //
00007 // Under terms of Contract DE-AC04-94AL85000, there is a non-exclusive
00008 // license for use of this work by or on behalf of the U.S. Government.
00009 //
00010 // Redistribution and use in source and binary forms, with or without
00011 // modification, are permitted provided that the following conditions are
00012 // met:
00013 //
00014 // 1. Redistributions of source code must retain the above copyright
00015 // notice, this list of conditions and the following disclaimer.
00016 //
00017 // 2. Redistributions in binary form must reproduce the above copyright
00018 // notice, this list of conditions and the following disclaimer in the
00019 // documentation and/or other materials provided with the distribution.
00020 //
00021 // 3. Neither the name of the Corporation nor the names of the
00022 // contributors may be used to endorse or promote products derived from
00023 // this software without specific prior written permission.
00024 //
00025 // THIS SOFTWARE IS PROVIDED BY SANDIA CORPORATION "AS IS" AND ANY
00026 // EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE
00027 // IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR
00028 // PURPOSE ARE DISCLAIMED. IN NO EVENT SHALL SANDIA CORPORATION OR THE
00029 // CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL,
00030 // EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO,
00031 // PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR
00032 // PROFITS; OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF
00033 // LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING
00034 // NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE OF THIS
00035 // SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
00036 //
00037 // Questions? Contact Michael A. Heroux (maherou@sandia.gov)
00038 //
00039 // ***********************************************************************
00040 // @HEADER
00041 
00042 #ifndef _TEUCHOS_SCALARTRAITS_CUDA_HPP_
00043 #define _TEUCHOS_SCALARTRAITS_CUDA_HPP_
00044 
00049 #include "Teuchos_ScalarTraitsDecl.hpp"
00050 #include "Teuchos_ScalarTraits.hpp"
00051 #include <cfloat>
00052 namespace Teuchos {
00053 
00054 #ifdef __CUDA_ARCH__
00055 template<>
00056 struct ScalarTraits<int>
00057 {
00058   typedef int magnitudeType;
00059   typedef int halfPrecision;
00060   typedef int doublePrecision;
00061   static const bool isComplex = false;
00062   static const bool isOrdinal = true;
00063   static const bool isComparable = true;
00064   static const bool hasMachineParameters = false;
00065   static inline __device__ __host__ magnitudeType magnitude(int a) { return (int)fabsf((float)a); }
00066   static inline __device__ __host__ int zero()  { return 0; }
00067   static inline __device__ __host__ int one()   { return 1; }
00068   static inline __device__ __host__ int conjugate(int x) { return x; }
00069   static inline __device__ __host__ int real(int x) { return x; }
00070   static inline __device__ __host__ int imag(int)   { return 0; }
00071   static inline __device__ __host__ bool isnaninf(int) { return false; }
00072   static inline __device__ __host__ int squareroot(int x) { return (int)sqrtf((float)x); }          // perhaps this cast should be replaced by an explicit call like __float2int_rn
00073   static inline __device__ __host__ int pow(int x, int y) { return (int)powf((float)x,(float)y); }  // perhaps this cast should be replaced by an explicit call like __float2int_rn
00074 
00075   // Dummy operations, need to exist for parsing when compiling everything with NVCC
00076   static inline __device__ __host__ int random() { return 9; }
00077   static inline __device__ __host__ void seedrandom(unsigned int ) {}
00078 };
00079 
00080 template<>
00081 struct ScalarTraits<unsigned int>
00082 {
00083   typedef unsigned int magnitudeType;
00084   typedef unsigned int halfPrecision;
00085   typedef unsigned int doublePrecision;
00086   static const bool isComplex = false;
00087   static const bool isOrdinal = true;
00088   static const bool isComparable = true;
00089   static const bool hasMachineParameters = false;
00090   static inline __device__ __host__ magnitudeType magnitude(unsigned int a) { return (unsigned int)fabsf((float)a); }
00091   static inline __device__ __host__ unsigned int zero()  { return 0; }
00092   static inline __device__ __host__ unsigned int one()   { return 1; }
00093   static inline __device__ __host__ unsigned int conjugate(unsigned int x) { return x; }
00094   static inline __device__ __host__ unsigned int real(unsigned int x) { return x; }
00095   static inline __device__ __host__ unsigned int imag(unsigned int)   { return 0; }
00096   static inline __device__ __host__ bool isnaninf(unsigned int) { return false; }
00097   static inline __device__ __host__ unsigned int squareroot(unsigned int x) { return (unsigned int)sqrtf((float)x); }          // perhaps this cast should be replaced by an explicit call like __float2int_rn
00098   static inline __device__ __host__ unsigned int pow(unsigned int x, unsigned int y) { return (unsigned int)powf((float)x,(float)y); }  // perhaps this cast should be replaced by an explicit call like __float2int_rn
00099 
00100   // Dummy operations, need to exist for parsing when compiling everything with NVCC
00101   static inline __device__ __host__ unsigned int random() { return 9; }
00102   static inline __device__ __host__ void seedrandom(unsigned int ) {}
00103 };
00104 
00105 template<>
00106 struct ScalarTraits<long int>
00107 {
00108   typedef long int magnitudeType;
00109   typedef long int halfPrecision;
00110   typedef long int doublePrecision;
00111   static const bool isComplex = false;
00112   static const bool isOrdinal = true;
00113   static const bool isComparable = true;
00114   static const bool hasMachineParameters = false;
00115   // Not defined: eps(), sfmin(), base(), prec(), t(), rnd(), emin(), rmin(), emax(), rmax()
00116   static inline __device__ __host__ magnitudeType magnitude(long int a) { return (long int)fabsf((float)a); }
00117   static inline __device__ __host__ long int zero()  { return 0; }
00118   static inline __device__ __host__ long int one()   { return 1; }
00119   static inline __device__ __host__ long int conjugate(long int x) { return x; }
00120   static inline __device__ __host__ long int real(long int x) { return x; }
00121   static inline __device__ __host__ long int imag(long int) { return 0; }
00122   static inline __device__ __host__ bool isnaninf(int) { return false; }
00123   static inline __device__ __host__ long int squareroot(long int x) { return (long int)sqrtf((float)x); }          // perhaps this cast should be replaced by an explicit call like __float2int_rn
00124   static inline __device__ __host__ long int pow(long int x, long int y) { return (long int)powf((float)x,(float)y); }  // perhaps this cast should be replaced by an explicit call like __float2int_rn
00125 
00126   // Dummy operations, need to exist for parsing when compiling everything with NVCC
00127   static inline __device__ __host__ long int random() { return 9; }
00128   static inline __device__ __host__ void seedrandom(unsigned int ) {}
00129 };
00130 
00131 template<>
00132 struct ScalarTraits<long unsigned int>
00133 {
00134   typedef long unsigned int magnitudeType;
00135   typedef long unsigned int halfPrecision;
00136   typedef long unsigned int doublePrecision;
00137   static const bool isComplex = false;
00138   static const bool isOrdinal = true;
00139   static const bool isComparable = true;
00140   static const bool hasMachineParameters = false;
00141   // Not defined: eps(), sfmin(), base(), prec(), t(), rnd(), emin(), rmin(), emax(), rmax()
00142   static inline __device__ __host__ magnitudeType magnitude(long unsigned int a) { return (long unsigned int)fabsf((float)a); }
00143   static inline __device__ __host__ long unsigned int zero()  { return 0; }
00144   static inline __device__ __host__ long unsigned int one()   { return 1; }
00145   static inline __device__ __host__ long unsigned int conjugate(long unsigned int x) { return x; }
00146   static inline __device__ __host__ long unsigned int real(long unsigned int x) { return x; }
00147   static inline __device__ __host__ long unsigned int imag(long unsigned int) { return 0; }
00148   static inline __device__ __host__ bool isnaninf(int) { return false; }
00149   static inline __device__ __host__ long unsigned int squareroot(long unsigned int x) { return (long unsigned int)sqrtf((float)x); }          // perhaps this cast should be replaced by an explicit call like __float2int_rn
00150   static inline __device__ __host__ long unsigned int pow(long unsigned int x, long unsigned int y) { return (long unsigned int)powf((float)x,(float)y); }  // perhaps this cast should be replaced by an explicit call like __float2int_rn
00151 
00152   // Dummy operations, need to exist for parsing when compiling everything with NVCC
00153   static inline __device__ __host__ long unsigned int random() { return 9; }
00154   static inline __device__ __host__ void seedrandom(unsigned int ) {}
00155 };
00156 
00157 template<>
00158 struct ScalarTraits<float>
00159 {
00160   typedef float magnitudeType;
00161   typedef float halfPrecision; // should become IEEE754-2008 binary16 or fp16 later, perhaps specified at configure according to architectural support
00162   typedef double doublePrecision;
00163   static const bool isComplex = false;
00164   static const bool isOrdinal = false;
00165   static const bool isComparable = true;
00166   static const bool hasMachineParameters = false;
00167   static inline __device__ __host__ magnitudeType magnitude(float a) { return fabsf(a); }
00168   static inline __device__ __host__ float zero()  { return(0.0f); }
00169   static inline __device__ __host__ float one()   { return(1.0f); }    
00170   static inline __device__ __host__ float conjugate(float x)   { return(x); }    
00171   static inline __device__ __host__ float real(float x) { return x; }
00172   static inline __device__ __host__ float imag(float)   { return zero(); }
00173   static inline __device__ __host__ bool  isnaninf(float x) { return isnan(x) || isinf(x); }
00174   static inline __device__ __host__ float squareroot(float x) { return sqrtf(x); }
00175   static inline __device__ __host__ float pow(float x, float y) { return powf(x,y); }
00176   static inline __device__ __host__ float eps() { return FLT_EPSILON; }
00177   static inline __device__ __host__ float t() { return FLT_MANT_DIG; }
00178   static inline __device__ __host__ float base() { return FLT_RADIX; }
00179   static inline __device__ __host__ float log10(float x ) { return ::log10f(x); }
00180 
00181   static inline __device__ __host__ float prec()  { return eps()*base(); }
00182   static inline __device__ __host__ float rnd()   { return 1.0f; }
00183   static inline __device__ __host__ float sfmin() { return FLT_MIN; }
00184   static inline __device__ __host__ float emin()  { return FLT_MIN_EXP; }
00185   static inline __device__ __host__ float rmin()  { return FLT_MIN; }
00186   static inline __device__ __host__ float emax()  { return FLT_MAX_EXP; }
00187   static inline __device__ __host__ float rmax()  { return FLT_MAX; }
00188   static inline __device__ __host__ float nan()   { return 0.0f/std::sin(0.0f); }
00189   static inline __device__ __host__ const char* name() { return "float"; }
00190 
00191   // Dummy operations, need to exist for parsing when compiling everything with NVCC
00192   static inline __device__ __host__ float random() { return 9.0f; }
00193   static inline __device__ __host__ void seedrandom(unsigned int ) {}
00194 };
00195 
00196 template<>
00197 struct ScalarTraits<double>
00198 {
00199   typedef double magnitudeType;
00200   typedef float  halfPrecision;
00201   typedef double doublePrecision;
00202   static const bool isComplex = false;
00203   static const bool isOrdinal = false;
00204   static const bool isComparable = true;
00205   static const bool hasMachineParameters = false;
00206   static inline __device__ __host__ magnitudeType magnitude(double a) { return abs(a); }
00207   static inline __device__ __host__ double zero()  { return(0.0); }
00208   static inline __device__ __host__ double one()   { return(1.0); }    
00209   static inline __device__ __host__ double conjugate(double x)   { return(x); }    
00210   static inline __device__ __host__ double real(double x) { return x; }
00211   static inline __device__ __host__ double imag(double)   { return zero(); }
00212   static inline __device__ __host__ bool  isnaninf(double x) { return isnan(x) || isinf(x); }
00213   static inline __device__ __host__ double squareroot(double x) { return sqrt(x); }
00214   static inline __device__ __host__ double pow(double x, double y) { return pow(x,y); }
00215   static inline __device__ __host__ double eps() { return DBL_EPSILON; }
00216   static inline __device__ __host__ double t() { return DBL_MANT_DIG; }
00217   static inline __device__ __host__ double base() { return FLT_RADIX; }
00218   static inline __device__ __host__ double log10(double x ) { return ::log10(x); }
00219 
00220   static inline __device__ __host__ double prec()  { return eps()*base(); }
00221   static inline __device__ __host__ double rnd()   { return 1.0; }
00222   static inline __device__ __host__ double sfmin() { return DBL_MIN; }
00223   static inline __device__ __host__ double emin()  { return DBL_MIN_EXP; }
00224   static inline __device__ __host__ double rmin()  { return DBL_MIN; }
00225   static inline __device__ __host__ double emax()  { return DBL_MAX_EXP; }
00226   static inline __device__ __host__ double rmax()  { return DBL_MAX; }
00227   static inline __device__ __host__ double nan()   { return 0.0/std::sin(0.0); }
00228   static inline __device__ __host__ const char* name() { return "double"; }
00229 
00230   // Dummy operations, need to exist for parsing when compiling everything with NVCC
00231   static inline __device__ __host__ double random() { return 9.0; }
00232   static inline __device__ __host__ void seedrandom(unsigned int ) {}
00233 };
00234 #endif // __CUDA_ARCH__
00235 
00236 } // Teuchos namespace
00237 
00238 #endif // _TEUCHOS_SCALARTRAITS_CUDA_HPP_
 All Classes Namespaces Files Functions Variables Typedefs Enumerations Enumerator Friends Defines