1 /* Minimal declarations for CUDA support. Testing purposes only. */ 2 3 #include <stddef.h> 4 5 #if __HIP__ || __CUDA__ 6 #define __constant__ __attribute__((constant)) 7 #define __device__ __attribute__((device)) 8 #define __global__ __attribute__((global)) 9 #define __host__ __attribute__((host)) 10 #define __shared__ __attribute__((shared)) 11 #if __HIP__ 12 #define __managed__ __attribute__((managed)) 13 #endif 14 #define __launch_bounds__(...) __attribute__((launch_bounds(__VA_ARGS__))) 15 #define __grid_constant__ __attribute__((grid_constant)) 16 #else 17 #define __constant__ 18 #define __device__ 19 #define __global__ 20 #define __host__ 21 #define __shared__ 22 #define __managed__ 23 #define __launch_bounds__(...) 24 #define __grid_constant__ 25 #endif 26 27 struct dim3 { 28 unsigned x, y, z; 29 __host__ __device__ dim3(unsigned x, unsigned y = 1, unsigned z = 1) : x(x), y(y), z(z) {} 30 }; 31 32 #if __HIP__ || HIP_PLATFORM 33 typedef struct hipStream *hipStream_t; 34 typedef enum hipError {} hipError_t; 35 int hipConfigureCall(dim3 gridSize, dim3 blockSize, size_t sharedSize = 0, 36 hipStream_t stream = 0); 37 extern "C" hipError_t __hipPushCallConfiguration(dim3 gridSize, dim3 blockSize, 38 size_t sharedSize = 0, 39 hipStream_t stream = 0); 40 #ifndef __HIP_API_PER_THREAD_DEFAULT_STREAM__ 41 extern "C" hipError_t hipLaunchKernel(const void *func, dim3 gridDim, 42 dim3 blockDim, void **args, 43 size_t sharedMem, 44 hipStream_t stream); 45 #else 46 extern "C" hipError_t hipLaunchKernel_spt(const void *func, dim3 gridDim, 47 dim3 blockDim, void **args, 48 size_t sharedMem, 49 hipStream_t stream); 50 #endif // __HIP_API_PER_THREAD_DEFAULT_STREAM__ 51 #elif __OFFLOAD_VIA_LLVM__ 52 extern "C" unsigned __llvmPushCallConfiguration(dim3 gridDim, dim3 blockDim, 53 size_t sharedMem = 0, void *stream = 0); 54 extern "C" unsigned llvmLaunchKernel(const void *func, dim3 gridDim, dim3 blockDim, 55 void **args, size_t sharedMem = 0, void *stream = 0); 56 #else 57 typedef struct cudaStream *cudaStream_t; 58 typedef enum cudaError {} cudaError_t; 59 extern "C" int cudaConfigureCall(dim3 gridSize, dim3 blockSize, 60 size_t sharedSize = 0, 61 cudaStream_t stream = 0); 62 extern "C" int __cudaPushCallConfiguration(dim3 gridSize, dim3 blockSize, 63 size_t sharedSize = 0, 64 cudaStream_t stream = 0); 65 extern "C" cudaError_t cudaLaunchKernel(const void *func, dim3 gridDim, 66 dim3 blockDim, void **args, 67 size_t sharedMem, cudaStream_t stream); 68 extern "C" cudaError_t cudaLaunchKernel_ptsz(const void *func, dim3 gridDim, 69 dim3 blockDim, void **args, 70 size_t sharedMem, cudaStream_t stream); 71 72 #endif 73 74 extern "C" __device__ int printf(const char*, ...); 75