libgomp/testsuite/libgomp.c-c++-common/icv-5.c

   1 /* { dg-do run } */
   2 /* { dg-set-target-env-var OMP_NUM_TEAMS_ALL "3" } */
   3 /* { dg-set-target-env-var OMP_NUM_TEAMS_DEV "4" } */
   4 /* { dg-set-target-env-var OMP_NUM_TEAMS "5" } */
   5 /* { dg-set-target-env-var OMP_NUM_TEAMS_DEV_0 "6" } */
   6 /* { dg-set-target-env-var OMP_NUM_TEAMS_DEV_1 "7" } */
   7 /* { dg-set-target-env-var OMP_NUM_TEAMS_DEV_2 "8" } */
   8 /* { dg-set-target-env-var OMP_TEAMS_THREAD_LIMIT_ALL "2" } */
   9 /* { dg-set-target-env-var OMP_TEAMS_THREAD_LIMIT_DEV "3" } */
  10 /* { dg-set-target-env-var OMP_TEAMS_THREAD_LIMIT "4" } */
  11 /* { dg-set-target-env-var OMP_TEAMS_THREAD_LIMIT_DEV_0 "5" } */
  12 /* { dg-set-target-env-var OMP_TEAMS_THREAD_LIMIT_DEV_1 "6" } */
  13 /* { dg-set-target-env-var OMP_TEAMS_THREAD_LIMIT_DEV_2 "7" } */
  14
  15 #include <omp.h>
  16 #include <stdlib.h>
  17 #include <unistd.h>
  18
  19 int
  20 main ()
  21 {
  22   if (omp_get_max_teams () != 5
  23       || omp_get_teams_thread_limit () != 4)
  24     abort ();
  25
  26   #pragma omp teams
  27   {
  28     if (omp_get_num_teams () > 5
  29         || omp_get_team_num () >= 5)
  30       abort ();
  31     #pragma omp parallel
  32     if (omp_get_thread_limit () > 4
  33         || omp_get_thread_num () >= 4)
  34       abort ();
  35   }
  36
  37   omp_set_num_teams (4);
  38   omp_set_teams_thread_limit (3);
  39   if (omp_get_max_teams () != 4
  40       || omp_get_teams_thread_limit () != 3)
  41     abort ();
  42
  43   #pragma omp teams
  44   {
  45     if (omp_get_num_teams () > 4
  46         || omp_get_team_num () >= 4)
  47       abort ();
  48     #pragma omp parallel
  49     if (omp_get_thread_limit () > 3
  50         || omp_get_thread_num () >= 3)
  51       abort ();
  52   }
  53
  54   #pragma omp teams num_teams(3) thread_limit(2)
  55   {
  56     if (omp_get_num_teams () != 3
  57         || omp_get_team_num () >= 3)
  58     abort ();
  59     #pragma omp parallel
  60     if (omp_get_thread_limit () > 2
  61         || omp_get_thread_num () >= 2)
  62       abort ();
  63   }
  64
  65   #pragma omp teams num_teams(5) thread_limit(4)
  66   {
  67     if (omp_get_num_teams () != 5
  68         || omp_get_team_num () >= 5)
  69     abort ();
  70     #pragma omp parallel
  71     if (omp_get_thread_limit () > 4
  72         || omp_get_thread_num () >= 4)
  73       abort ();
  74   }
  75
  76   int num_devices = omp_get_num_devices () > 3 ? 3 : omp_get_num_devices ();
  77
  78   for (int i = 0; i < num_devices; i++)
  79     {
  80       #pragma omp target device (i)
  81       if (omp_get_max_teams () != 6 + i
  82           || omp_get_teams_thread_limit () != 5 + i)
  83         abort ();
  84
  85       #pragma omp target device (i)
  86       #pragma omp teams
  87       #pragma omp parallel
  88       if (omp_get_thread_limit () > 5 + i
  89           || omp_get_thread_num () >= 5 + i)
  90         abort ();
  91
  92       #pragma omp target device (i)
  93       {
  94         omp_set_num_teams (5 + i);
  95         omp_set_teams_thread_limit (4 + i);
  96         if (omp_get_max_teams () != 5 + i
  97             || omp_get_teams_thread_limit () != 4 + i)
  98           abort ();
  99       }
 100
 101       /* omp_set_num_teams and omp_set_teams_thread_limit above set the value
 102          of nteams-var and teams-thread-limit-var ICVs on device 'i', which has
 103          scope 'device' and should be avaible in subsequent target regions.  */
 104       #pragma omp target device (i)
 105       if (omp_get_max_teams () != 5 + i
 106           || omp_get_teams_thread_limit () != 4 + i)
 107         abort ();
 108
 109       #pragma omp target device (i)
 110       #pragma omp teams
 111       {
 112         if (omp_get_num_teams () > 5 + i
 113             || omp_get_team_num () >= 5 + i)
 114           abort ();
 115         #pragma omp parallel
 116         if (omp_get_thread_limit () > 4 + i
 117             || omp_get_thread_num () >= 4 + i)
 118           abort ();
 119       }
 120
 121       #pragma omp target device (i)
 122       #pragma omp teams num_teams(6 + i) thread_limit(5 + i)
 123       {
 124         if (omp_get_num_teams () > 6 + i
 125             || omp_get_team_num () >= 6 + i)
 126           abort ();
 127         #pragma omp parallel
 128         if (omp_get_thread_limit () > 5 + i
 129             || omp_get_thread_num () >= 5 + i
 130             || omp_get_num_teams () > 6 + i
 131             || omp_get_team_num () >= 6 + i)
 132           abort ();
 133       }
 134
 135       #pragma omp target device (i)
 136       #pragma omp teams num_teams(4 + i) thread_limit(3 + i)
 137       {
 138         if (omp_get_num_teams () > 4 + i
 139             || omp_get_team_num () >= 4 + i)
 140           abort ();
 141         #pragma omp parallel
 142         if (omp_get_thread_limit () > 3 + i
 143             || omp_get_thread_num () >= 3 + i
 144             || omp_get_num_teams () > 4 + i
 145             || omp_get_team_num () >= 4 + i)
 146           abort ();
 147       }
 148
 149       #pragma omp target device (i)
 150       #pragma omp teams thread_limit(3 + i) num_teams(4 + i)
 151       {
 152         if (omp_get_num_teams () > 4 + i
 153             || omp_get_team_num () >= 4 + i)
 154           abort ();
 155         #pragma omp parallel
 156         if (omp_get_thread_limit () > 3 + i
 157             || omp_get_thread_num () >= 3 + i
 158             || omp_get_num_teams () > 4 + i
 159             || omp_get_team_num () >= 4 + i)
 160           abort ();
 161       }
 162
 163       /* The NUM_TEAMS and THREAD_LIMIT clauses should not change the values
 164          of the corresponding ICVs.  */
 165       #pragma omp target device (i)
 166       if (omp_get_max_teams () != 5 + i
 167           || omp_get_teams_thread_limit () != 4 + i)
 168         abort ();
 169
 170       /* This tests a large number of teams and threads.  If it is larger than
 171          2^15+1 then the according argument in the kernels arguments list
 172          is encoded with two items instead of one.  */
 173       intptr_t large_num_teams = 66000;
 174       intptr_t large_threads_limit = 67000;
 175       #pragma omp target device (i)
 176       {
 177         omp_set_num_teams (large_num_teams + i);
 178         omp_set_teams_thread_limit (large_threads_limit + i);
 179         if (omp_get_max_teams () != large_num_teams + i
 180             || omp_get_teams_thread_limit () != large_threads_limit + i)
 181           abort ();
 182       }
 183
 184       #pragma omp target device (i)
 185         if (omp_get_max_teams () != large_num_teams + i
 186             || omp_get_teams_thread_limit () != large_threads_limit + i)
 187         abort ();
 188
 189       #pragma omp target device (i)
 190       #pragma omp teams
 191       {
 192         if (omp_get_num_teams () > large_num_teams + i
 193             || omp_get_team_num () >= large_num_teams + i)
 194           abort ();
 195         #pragma omp parallel
 196         if (omp_get_thread_limit () > large_threads_limit + i
 197             || omp_get_thread_num () >= large_threads_limit + i)
 198           abort ();
 199       }
 200     }
 201
 202   return 0;
 203 }