280 lines
12 KiB
Plaintext
280 lines
12 KiB
Plaintext
|
#pragma once
|
||
|
|
|
||
|
|
#include <cuda.h>
|
||
|
|
#include <cuda_runtime.h>
|
||
|
|
#include <float.h>
|
||
|
|
#include <limits.h>
|
||
|
|
|
||
|
|
#include "cub/cub.cuh"
|
||
|
|
|
||
|
|
#define checkCuda(expression) \
|
||
|
|
{ \
|
||
|
|
cudaError_t status = (expression); \
|
||
|
|
if (status != cudaSuccess) { \
|
||
|
|
printf("CUDA Runtime Error: %s at %s:%d\n", cudaGetErrorString(expression), __FILE__, __LINE__); \
|
||
|
|
std::exit(EXIT_FAILURE); \
|
||
|
|
} \
|
||
|
|
}
|
||
|
|
|
||
|
|
#define checkCurand(expression) \
|
||
|
|
{ \
|
||
|
|
curandStatus_t status = (expression); \
|
||
|
|
if (status != CURAND_STATUS_SUCCESS) { \
|
||
|
|
printf("Curand Error at %s:%d\n", __FILE__, __LINE__); \
|
||
|
|
std::exit(EXIT_FAILURE); \
|
||
|
|
} \
|
||
|
|
}
|
||
|
|
|
||
|
|
#define allocateCuda(var, size, type) \
|
||
|
|
{ \
|
||
|
|
cudaError_t status = cudaMalloc(&(var), (size) * sizeof(type)); \
|
||
|
|
if (status != cudaSuccess) { \
|
||
|
|
printf("cudaMalloc failed for " #var " at %s:%d\n", __FILE__, __LINE__); \
|
||
|
|
} \
|
||
|
|
}
|
||
|
|
|
||
|
|
#define allocateCopyCuda(var, rhs, size) \
|
||
|
|
{ \
|
||
|
|
allocateCuda(var, size, decltype(*rhs)); \
|
||
|
|
checkCuda(cudaMemcpy(var, rhs, sizeof(decltype(*rhs)) * (size), cudaMemcpyHostToDevice)); \
|
||
|
|
}
|
||
|
|
|
||
|
|
#define allocateCopyCpu(var, rhs, size, T) \
|
||
|
|
{ \
|
||
|
|
var = (T*)malloc(sizeof(T) * (size)); \
|
||
|
|
checkCuda(cudaMemcpy((void*)var, (void*)rhs, sizeof(T) * (size), cudaMemcpyDeviceToHost)); \
|
||
|
|
}
|
||
|
|
|
||
|
|
|
||
|
|
// For cuda::numeric_limits
|
||
|
|
namespace cuda { // namespace cuda
|
||
|
|
|
||
|
|
template <typename T>
|
||
|
|
struct numeric_limits_base {
|
||
|
|
typedef T type;
|
||
|
|
};
|
||
|
|
template <typename T>
|
||
|
|
struct numeric_limits : public numeric_limits_base<T> {};
|
||
|
|
|
||
|
|
template <>
|
||
|
|
struct numeric_limits<char> : public numeric_limits_base<char> {
|
||
|
|
/** The minimum finite value, or for floating types with
|
||
|
|
denormalization, the minimum positive normalized value. */
|
||
|
|
__host__ __device__ static constexpr type min() noexcept { return CHAR_MIN; }
|
||
|
|
|
||
|
|
/** The maximum finite value. */
|
||
|
|
__host__ __device__ static constexpr type max() noexcept { return CHAR_MAX; }
|
||
|
|
|
||
|
|
/** A finite value x such that there is no other finite value y
|
||
|
|
* where y < x. */
|
||
|
|
__host__ __device__ static constexpr type lowest() noexcept { return CHAR_MIN; }
|
||
|
|
|
||
|
|
/** A the machine epsilon. */
|
||
|
|
__host__ __device__ static constexpr type epsilon() noexcept { return 0; }
|
||
|
|
};
|
||
|
|
|
||
|
|
template <>
|
||
|
|
struct numeric_limits<unsigned char> : public numeric_limits_base<unsigned char> {
|
||
|
|
/** The minimum finite value, or for floating types with
|
||
|
|
denormalization, the minimum positive normalized value. */
|
||
|
|
__host__ __device__ static constexpr type min() noexcept { return 0; }
|
||
|
|
|
||
|
|
/** The maximum finite value. */
|
||
|
|
__host__ __device__ static constexpr type max() noexcept { return UCHAR_MAX; }
|
||
|
|
|
||
|
|
/** A finite value x such that there is no other finite value y
|
||
|
|
* where y < x. */
|
||
|
|
__host__ __device__ static constexpr type lowest() noexcept { return 0; }
|
||
|
|
|
||
|
|
/** A the machine epsilon. */
|
||
|
|
__host__ __device__ static constexpr type epsilon() noexcept { return 0; }
|
||
|
|
};
|
||
|
|
|
||
|
|
template <>
|
||
|
|
struct numeric_limits<short> : public numeric_limits_base<short> {
|
||
|
|
/** The minimum finite value, or for floating types with
|
||
|
|
denormalization, the minimum positive normalized value. */
|
||
|
|
__host__ __device__ static constexpr type min() noexcept { return SHRT_MIN; }
|
||
|
|
|
||
|
|
/** The maximum finite value. */
|
||
|
|
__host__ __device__ static constexpr type max() noexcept { return SHRT_MAX; }
|
||
|
|
|
||
|
|
/** A finite value x such that there is no other finite value y
|
||
|
|
* where y < x. */
|
||
|
|
__host__ __device__ static constexpr type lowest() noexcept { return SHRT_MIN; }
|
||
|
|
|
||
|
|
/** A the machine epsilon. */
|
||
|
|
__host__ __device__ static constexpr type epsilon() noexcept { return 0; }
|
||
|
|
};
|
||
|
|
|
||
|
|
template <>
|
||
|
|
struct numeric_limits<unsigned short> : public numeric_limits_base<unsigned short> {
|
||
|
|
/** The minimum finite value, or for floating types with
|
||
|
|
denormalization, the minimum positive normalized value. */
|
||
|
|
__host__ __device__ static constexpr type min() noexcept { return 0; }
|
||
|
|
|
||
|
|
/** The maximum finite value. */
|
||
|
|
__host__ __device__ static constexpr type max() noexcept { return USHRT_MAX; }
|
||
|
|
|
||
|
|
/** A finite value x such that there is no other finite value y
|
||
|
|
* where y < x. */
|
||
|
|
__host__ __device__ static constexpr type lowest() noexcept { return 0; }
|
||
|
|
|
||
|
|
/** A the machine epsilon. */
|
||
|
|
__host__ __device__ static constexpr type epsilon() noexcept { return 0; }
|
||
|
|
};
|
||
|
|
|
||
|
|
template <>
|
||
|
|
struct numeric_limits<int> : public numeric_limits_base<int> {
|
||
|
|
/** The minimum finite value, or for floating types with
|
||
|
|
denormalization, the minimum positive normalized value. */
|
||
|
|
__host__ __device__ static constexpr type min() noexcept { return INT_MIN; }
|
||
|
|
|
||
|
|
/** The maximum finite value. */
|
||
|
|
__host__ __device__ static constexpr type max() noexcept { return INT_MAX; }
|
||
|
|
|
||
|
|
/** A finite value x such that there is no other finite value y
|
||
|
|
* where y < x. */
|
||
|
|
__host__ __device__ static constexpr type lowest() noexcept { return INT_MIN; }
|
||
|
|
|
||
|
|
/** A the machine epsilon. */
|
||
|
|
__host__ __device__ static constexpr type epsilon() noexcept { return 0; }
|
||
|
|
};
|
||
|
|
|
||
|
|
template <>
|
||
|
|
struct numeric_limits<unsigned int> : public numeric_limits_base<unsigned int> {
|
||
|
|
/** The minimum finite value, or for floating types with
|
||
|
|
denormalization, the minimum positive normalized value. */
|
||
|
|
__host__ __device__ static constexpr type min() noexcept { return 0; }
|
||
|
|
|
||
|
|
/** The maximum finite value. */
|
||
|
|
__host__ __device__ static constexpr type max() noexcept { return UINT_MAX; }
|
||
|
|
|
||
|
|
/** A finite value x such that there is no other finite value y
|
||
|
|
* where y < x. */
|
||
|
|
__host__ __device__ static constexpr type lowest() noexcept { return 0; }
|
||
|
|
|
||
|
|
/** A the machine epsilon. */
|
||
|
|
__host__ __device__ static constexpr type epsilon() noexcept { return 0; }
|
||
|
|
};
|
||
|
|
|
||
|
|
template <>
|
||
|
|
struct numeric_limits<long> : public numeric_limits_base<long> {
|
||
|
|
/** The minimum finite value, or for floating types with
|
||
|
|
denormalization, the minimum positive normalized value. */
|
||
|
|
__host__ __device__ static constexpr type min() noexcept { return LONG_MIN; }
|
||
|
|
|
||
|
|
/** The maximum finite value. */
|
||
|
|
__host__ __device__ static constexpr type max() noexcept { return LONG_MAX; }
|
||
|
|
|
||
|
|
/** A finite value x such that there is no other finite value y
|
||
|
|
* where y < x. */
|
||
|
|
__host__ __device__ static constexpr type lowest() noexcept { return LONG_MIN; }
|
||
|
|
|
||
|
|
/** A the machine epsilon. */
|
||
|
|
__host__ __device__ static constexpr type epsilon() noexcept { return 0; }
|
||
|
|
};
|
||
|
|
|
||
|
|
template <>
|
||
|
|
struct numeric_limits<unsigned long> : public numeric_limits_base<unsigned long> {
|
||
|
|
/** The minimum finite value, or for floating types with
|
||
|
|
denormalization, the minimum positive normalized value. */
|
||
|
|
__host__ __device__ static constexpr type min() noexcept { return 0; }
|
||
|
|
|
||
|
|
/** The maximum finite value. */
|
||
|
|
__host__ __device__ static constexpr type max() noexcept { return ULONG_MAX; }
|
||
|
|
|
||
|
|
/** A finite value x such that there is no other finite value y
|
||
|
|
* where y < x. */
|
||
|
|
__host__ __device__ static constexpr type lowest() noexcept { return 0; }
|
||
|
|
|
||
|
|
/** A the machine epsilon. */
|
||
|
|
__host__ __device__ static constexpr type epsilon() noexcept { return 0; }
|
||
|
|
};
|
||
|
|
|
||
|
|
template <>
|
||
|
|
struct numeric_limits<long long> : public numeric_limits_base<long long> {
|
||
|
|
/** The minimum finite value, or for floating types with
|
||
|
|
denormalization, the minimum positive normalized value. */
|
||
|
|
__host__ __device__ static constexpr type min() noexcept { return LLONG_MIN; }
|
||
|
|
|
||
|
|
/** The maximum finite value. */
|
||
|
|
__host__ __device__ static constexpr type max() noexcept { return LLONG_MAX; }
|
||
|
|
|
||
|
|
/** A finite value x such that there is no other finite value y
|
||
|
|
* where y < x. */
|
||
|
|
__host__ __device__ static constexpr type lowest() noexcept { return LLONG_MIN; }
|
||
|
|
|
||
|
|
/** A the machine epsilon. */
|
||
|
|
__host__ __device__ static constexpr type epsilon() noexcept { return 0; }
|
||
|
|
};
|
||
|
|
|
||
|
|
template <>
|
||
|
|
struct numeric_limits<unsigned long long> : public numeric_limits_base<unsigned long long> {
|
||
|
|
/** The minimum finite value, or for floating types with
|
||
|
|
denormalization, the minimum positive normalized value. */
|
||
|
|
__host__ __device__ static constexpr type min() noexcept { return 0; }
|
||
|
|
|
||
|
|
/** The maximum finite value. */
|
||
|
|
__host__ __device__ static constexpr type max() noexcept { return ULLONG_MAX; }
|
||
|
|
|
||
|
|
/** A finite value x such that there is no other finite value y
|
||
|
|
* where y < x. */
|
||
|
|
__host__ __device__ static constexpr type lowest() noexcept { return 0; }
|
||
|
|
|
||
|
|
/** A the machine epsilon. */
|
||
|
|
__host__ __device__ static constexpr type epsilon() noexcept { return 0; }
|
||
|
|
};
|
||
|
|
|
||
|
|
template <>
|
||
|
|
struct numeric_limits<float> : public numeric_limits_base<float> {
|
||
|
|
/** The minimum finite value, or for floating types with
|
||
|
|
denormalization, the minimum positive normalized value. */
|
||
|
|
__host__ __device__ static constexpr type min() noexcept { return FLT_MIN; }
|
||
|
|
|
||
|
|
/** The maximum finite value. */
|
||
|
|
__host__ __device__ static constexpr type max() noexcept { return FLT_MAX; }
|
||
|
|
|
||
|
|
/** A finite value x such that there is no other finite value y
|
||
|
|
* where y < x. */
|
||
|
|
__host__ __device__ static constexpr type lowest() noexcept { return -FLT_MAX; }
|
||
|
|
|
||
|
|
/** A the machine epsilon. */
|
||
|
|
__host__ __device__ static constexpr type epsilon() noexcept { return FLT_EPSILON; }
|
||
|
|
};
|
||
|
|
|
||
|
|
template <>
|
||
|
|
struct numeric_limits<double> : public numeric_limits_base<double> {
|
||
|
|
/** The minimum finite value, or for floating types with
|
||
|
|
denormalization, the minimum positive normalized value. */
|
||
|
|
__host__ __device__ static constexpr type min() noexcept { return DBL_MIN; }
|
||
|
|
|
||
|
|
/** The maximum finite value. */
|
||
|
|
__host__ __device__ static constexpr type max() noexcept { return DBL_MAX; }
|
||
|
|
|
||
|
|
/** A finite value x such that there is no other finite value y
|
||
|
|
* where y < x. */
|
||
|
|
__host__ __device__ static constexpr type lowest() noexcept { return -DBL_MAX; }
|
||
|
|
|
||
|
|
/** A the machine epsilon. */
|
||
|
|
__host__ __device__ static constexpr type epsilon() noexcept { return DBL_EPSILON; }
|
||
|
|
};
|
||
|
|
|
||
|
|
template <>
|
||
|
|
struct numeric_limits<long double> : public numeric_limits_base<long double> {
|
||
|
|
/** The minimum finite value, or for floating types with
|
||
|
|
denormalization, the minimum positive normalized value. */
|
||
|
|
__host__ __device__ static constexpr type min() noexcept { return LDBL_MIN; }
|
||
|
|
|
||
|
|
/** The maximum finite value. */
|
||
|
|
__host__ __device__ static constexpr type max() noexcept { return LDBL_MAX; }
|
||
|
|
|
||
|
|
/** A finite value x such that there is no other finite value y
|
||
|
|
* where y < x. */
|
||
|
|
__host__ __device__ static constexpr type lowest() noexcept { return -LDBL_MAX; }
|
||
|
|
|
||
|
|
/** A the machine epsilon. */
|
||
|
|
__host__ __device__ static constexpr type epsilon() noexcept { return LDBL_EPSILON; }
|
||
|
|
};
|
||
|
|
} // namespace cuda
|