11void gpu_memory_pool_alloc(
void **p,
size_t req_size);
12void gpu_memory_pool_free(
void *ptr);
13void gpu_memory_pool_purge();
14void gpu_memory_pool_report();
21void seed_device_rng(
unsigned long long seed);
39#include <cuda_runtime.h>
42using gpuError = cudaError;
43#define gpuSuccess cudaSuccess
48#define gpuMalloc(a, b) gpu_memory_pool_alloc((void **)a, b)
49#define gpuFree(a) gpu_memory_pool_free(a)
50#define gpuMemPoolPurge() gpu_memory_pool_purge()
51#define gpuMemPoolReport() gpu_memory_pool_report()
57#define gpuMemPoolPurge() do { } while (0)
58#define gpuMemPoolReport() do { } while (0)
61#ifdef CUDA_MALLOC_ASYNC
62#define gpuMalloc(a, b) GPU_CHECK(cudaMallocAsync(a, b, 0))
63#define gpuFree(a) GPU_CHECK(cudaFreeAsync(a, 0))
66#define gpuMalloc(a, b) GPU_CHECK(cudaMalloc((void **)a, b))
67#define gpuFree(a) GPU_CHECK(cudaFree(a))
75#define gpuGetLastError cudaGetLastError
76#define gpuMemcpy(a, b, c, d) GPU_CHECK(cudaMemcpy(a, b, c, d))
77#define gpuMemcpyHostToDevice cudaMemcpyHostToDevice
78#define gpuMemcpyDeviceToHost cudaMemcpyDeviceToHost
79#define gpuMemcpyDeviceToDevice cudaMemcpyDeviceToDevice
80#define gpuDeviceSynchronize() GPU_CHECK(cudaDeviceSynchronize())
81#define gpuStreamSynchronize(a) GPU_CHECK(cudaStreamSynchronize(a))
82#define gpuMemset(a,b,c) GPU_CHECK(cudaMemset(a,b,c))
83#define gpuMemcpyToSymbol(a, b, size, c, dir) GPU_CHECK(cudaMemcpyToSymbol(a, b, size, c, dir))
85#define GPUTYPESTR "CUDA"
88#define _GPU_DEVICE_COMPILE_ __CUDA_ARCH__
96#include <hip/hip_runtime.h>
97#include <hiprand/hiprand.h>
101using gpuError = hipError_t;
102#define gpuSuccess hipSuccess
106#ifdef GPU_MEMORY_POOL
107#define gpuMalloc(a, b) gpu_memory_pool_alloc((void **)a, b)
108#define gpuFree(a) gpu_memory_pool_free(a)
109#define gpuMemPoolPurge() gpu_memory_pool_purge()
110#define gpuMemPoolReport() gpu_memory_pool_report()
117#define gpuMemPoolPurge() do {} while (0)
118#define gpuMemPoolReport() do {} while (0)
121#define gpuMalloc(a, b) GPU_CHECK(hipMalloc((void **)a, b))
122#define gpuFree(a) GPU_CHECK(hipFree(a))
126#define gpuGetLastError hipGetLastError
127#define gpuMemcpy(a, b, siz, d) GPU_CHECK(hipMemcpy(a, b, siz, d))
128#define gpuMemcpyHostToDevice hipMemcpyHostToDevice
129#define gpuMemcpyDeviceToHost hipMemcpyDeviceToHost
130#define gpuMemcpyDeviceToDevice hipMemcpyDeviceToDevice
131#define gpuDeviceSynchronize() GPU_CHECK(hipDeviceSynchronize())
132#define gpuStreamSynchronize(a) GPU_CHECK(hipStreamSynchronize(a))
133#define gpuMemset(a,b,c) GPU_CHECK(hipMemset(a,b,c))
134#define gpuMemcpyToSymbol(a, b, size, c, dir) \
135 GPU_CHECK(hipMemcpyToSymbol(HIP_SYMBOL(a), b, size, c, dir))
138#define GPUTYPESTR "HIP"
140#ifdef __HIP_DEVICE_COMPILE__
141#define _GPU_DEVICE_COMPILE_ __HIP_DEVICE_COMPILE__
150#define GPU_CHECK(cmd) \
153 gpu_exit_on_error(code, #cmd, __FILE__, __LINE__); \
156#define check_device_error(msg) gpu_exit_on_error(msg, __FILE__, __LINE__)
157#define check_device_error_code(code, msg) \
158 gpu_exit_on_error(code, msg, __FILE__, __LINE__)
159void gpu_exit_on_error(
const char *msg,
const char *file,
int line);
160void gpu_exit_on_error(gpuError code,
const char *msg,
const char *file,
int line);
163inline void synchronize_threads() {
164 gpuDeviceSynchronize();
179#define gpuMalloc(a, b) do {} while(0)
180#define gpuFree(a) do {} while(0)
181#define gpuMemcpy(a, b, siz, d) do {} while(0)
182#define gpuMemcpyHostToDevice 1
183#define gpuMemcpyDeviceToHost 2
184#define gpuMemset(a,b,c) do {} while(0)
185#define gpuMemcpyToSymbol(a, b, size, c, dir) do {} while(0)
187#define gpuMemPoolPurge() do {} while(0)
188#define gpuMemPoolReport() do {} while(0)
190#define check_device_error(msg) do {} while(0)
191#define check_device_error_code(code, msg) do {} while(0)
193#define gpuStreamSynchronize(a) do {} while(0)
194#define gpuDeviceSynchronize() do {} while(0)
196#define gpuGetLastError cudaGetLastError
202#define GPUTYPESTR "NONE"
205inline void synchronize_threads() {}
211void initialize_gpu(
int rank,
int device);
212void gpu_device_info();
223struct is_arithmetic : std::integral_constant<bool, std::is_arithmetic<T>::value> {};
225template <
class T,
class U>
226struct is_assignable : std::integral_constant<bool, std::is_assignable<T, U>::value> {};
229struct is_floating_point
230 : std::integral_constant<bool, std::is_floating_point<T>::value> {};
Implement hila::swap for gauge fields.
void free_device_rng()
Free GPU RNG state, does nothing on non-GPU archs.