Last active
March 16, 2018 15:28
-
-
Save fwyzard/2b20dd522f2716ec363ea6f1e30c56b4 to your computer and use it in GitHub Desktop.
Simple file showing how __host__ __device_ function interact
This file contains bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
#include <cstdio> | |
#include <cuda.h> | |
__host__ __device__ | |
void where() { | |
#if defined __CUDA_ARCH__ | |
printf(" on the device"); | |
#else | |
printf(" on the host"); | |
#endif | |
} | |
struct HostOnlyType { | |
__host__ /* __device__ */ | |
void call() | |
{ | |
printf("__host__ function"); | |
where(); | |
printf("\n"); | |
} | |
}; | |
struct DeviceOnlyType { | |
/* __host__ */ __device__ | |
void call() | |
{ | |
printf("__device__ function"); | |
where(); | |
printf("\n"); | |
} | |
}; | |
struct HostDeviceType { | |
__host__ __device__ | |
void call() | |
{ | |
printf("__host__ __device__ function"); | |
where(); | |
printf("\n"); | |
} | |
}; | |
template <typename T> | |
struct Dispatcher { | |
__host__ // only if T::call() is a __host__ function ? | |
__device__ // only if T::call() is a __device__ function ? | |
static void dispatch() { | |
T t; | |
t.call(); | |
} | |
}; | |
template <typename T> | |
__global__ | |
void kernel(void) { | |
Dispatcher<T>::dispatch(); | |
} | |
int main(void) { | |
Dispatcher<HostOnlyType>::dispatch(); | |
//Dispatcher<DeviceOnlyType>::dispatch(); | |
Dispatcher<HostDeviceType>::dispatch(); | |
//kernel<HostOnlyType><<<1,1>>>(); | |
kernel<DeviceOnlyType><<<1,1>>>(); | |
kernel<HostDeviceType><<<1,1>>>(); | |
cudaDeviceSynchronize(); | |
} |
Sign up for free
to join this conversation on GitHub.
Already have an account?
Sign in to comment