51 lines · c
1// Check that omp atomic is permitted and behaves when strictly nested within2// omp target teams. This is an extension to OpenMP 5.2 and is enabled by3// default.4 5// RUN: %libomptarget-compile-run-and-check-generic6 7#include <omp.h>8#include <stdbool.h>9#include <stdio.h>10#include <string.h>11 12// High parallelism increases our chances of detecting a lack of atomicity.13#define NUM_TEAMS_TRY 25614 15int main() {16 // CHECK: update: num_teams=[[#NUM_TEAMS:]]{{$}}17 // CHECK-NEXT: update: x=[[#NUM_TEAMS]]{{$}}18 int x = 0;19 int numTeams;20#pragma omp target teams num_teams(NUM_TEAMS_TRY) map(tofrom : x, numTeams)21 {22#pragma omp atomic update23 ++x;24 if (omp_get_team_num() == 0)25 numTeams = omp_get_num_teams();26 }27 printf("update: num_teams=%d\n", numTeams);28 printf("update: x=%d\n", x);29 30 // CHECK-NEXT: capture: x=[[#NUM_TEAMS]]{{$}}31 // CHECK-NEXT: capture: xCapturedCount=[[#NUM_TEAMS]]{{$}}32 bool xCaptured[numTeams];33 memset(xCaptured, 0, sizeof xCaptured);34 x = 0;35#pragma omp target teams num_teams(NUM_TEAMS_TRY) map(tofrom : x, numTeams)36 {37 int v;38#pragma omp atomic capture39 v = x++;40 xCaptured[v] = true;41 }42 printf("capture: x=%d\n", x);43 int xCapturedCount = 0;44 for (int i = 0; i < numTeams; ++i) {45 if (xCaptured[i])46 ++xCapturedCount;47 }48 printf("capture: xCapturedCount=%d\n", xCapturedCount);49 return 0;50}51