on_device_arch.h 1.1 KB

123456789101112131415161718192021222324252627282930313233343536373839404142434445464748495051525354555657585960616263646566
  1. #include <omp.h>
  2. #include <gomp-constants.h>
  3. /* static */ int
  4. device_arch_nvptx (void)
  5. {
  6. return GOMP_DEVICE_NVIDIA_PTX;
  7. }
  8. /* static */ int
  9. device_arch_intel_mic (void)
  10. {
  11. return GOMP_DEVICE_INTEL_MIC;
  12. }
  13. #pragma omp declare variant (device_arch_nvptx) match(construct={target},device={arch(nvptx)})
  14. #pragma omp declare variant (device_arch_intel_mic) match(construct={target},device={arch(intel_mic)})
  15. /* static */ int
  16. device_arch (void)
  17. {
  18. return GOMP_DEVICE_DEFAULT;
  19. }
  20. static int
  21. on_device_arch (int d)
  22. {
  23. int d_cur;
  24. #pragma omp target map(from:d_cur)
  25. d_cur = device_arch ();
  26. return d_cur == d;
  27. }
  28. int
  29. on_device_arch_nvptx ()
  30. {
  31. return on_device_arch (GOMP_DEVICE_NVIDIA_PTX);
  32. }
  33. int
  34. on_device_arch_intel_mic ()
  35. {
  36. return on_device_arch (GOMP_DEVICE_INTEL_MIC);
  37. }
  38. static int
  39. any_device_arch (int d)
  40. {
  41. int nd = omp_get_num_devices ();
  42. for (int i = 0; i < nd; ++i)
  43. {
  44. int d_cur;
  45. #pragma omp target device(i) map(from:d_cur)
  46. d_cur = device_arch ();
  47. if (d_cur == d)
  48. return 1;
  49. }
  50. return 0;
  51. }
  52. int
  53. any_device_arch_intel_mic ()
  54. {
  55. return any_device_arch (GOMP_DEVICE_INTEL_MIC);
  56. }