124f836e8SJoachim Protze // RUN: %libomptarget-compilexx-run-and-check-generic 26aa7228aSAlexey Bataev 3fb9fc3c9SJon Chesterfield // Error on the gpu that crashes the host 4034adaf5SJoseph Huber // UNSUPPORTED: amdgcn-amd-amdhsa 5ae23be84SJoseph Huber // UNSUPPORTED: amdgcn-amd-amdhsa-oldDriver 6*d5d83663SJoseph Huber // UNSUPPORTED: amdgcn-amd-amdhsa-LTO 7fb9fc3c9SJon Chesterfield 86aa7228aSAlexey Bataev #include <iostream> 96aa7228aSAlexey Bataev 106aa7228aSAlexey Bataev template <typename LOOP_BODY> forall(int Begin,int End,LOOP_BODY LoopBody)116aa7228aSAlexey Bataevinline void forall(int Begin, int End, LOOP_BODY LoopBody) { 126aa7228aSAlexey Bataev #pragma omp target parallel for schedule(static) 136aa7228aSAlexey Bataev for (int I = Begin; I < End; ++I) { 146aa7228aSAlexey Bataev LoopBody(I); 156aa7228aSAlexey Bataev } 166aa7228aSAlexey Bataev } 176aa7228aSAlexey Bataev 186aa7228aSAlexey Bataev #define N (1000) 196aa7228aSAlexey Bataev 206aa7228aSAlexey Bataev // 216aa7228aSAlexey Bataev // Demonstration of the RAJA abstraction using lambdas 226aa7228aSAlexey Bataev // Requires data mapping onto the target section 236aa7228aSAlexey Bataev // main()246aa7228aSAlexey Bataevint main() { 256aa7228aSAlexey Bataev double A[N], B[N], C[N]; 266aa7228aSAlexey Bataev 276aa7228aSAlexey Bataev for (int I = 0; I < N; I++) { 286aa7228aSAlexey Bataev A[I] = I + 1; 296aa7228aSAlexey Bataev B[I] = -I; 306aa7228aSAlexey Bataev C[I] = -9; 316aa7228aSAlexey Bataev } 326aa7228aSAlexey Bataev 336aa7228aSAlexey Bataev #pragma omp target data map(tofrom : C [0:N]) map(to : A [0:N], B [0:N]) 346aa7228aSAlexey Bataev { 356aa7228aSAlexey Bataev forall(0, N, [&](int I) { C[I] += A[I] + B[I]; }); 366aa7228aSAlexey Bataev } 376aa7228aSAlexey Bataev 386aa7228aSAlexey Bataev int Fail = 0; 396aa7228aSAlexey Bataev for (int I = 0; I < N; I++) { 406aa7228aSAlexey Bataev if (C[I] != -8) { 416aa7228aSAlexey Bataev std::cout << "Failed at " << I << " with val " << C[I] << std::endl; 426aa7228aSAlexey Bataev Fail = 1; 436aa7228aSAlexey Bataev } 446aa7228aSAlexey Bataev } 456aa7228aSAlexey Bataev 466aa7228aSAlexey Bataev // CHECK: Succeeded 476aa7228aSAlexey Bataev if (Fail) { 486aa7228aSAlexey Bataev std::cout << "Failed" << std::endl; 496aa7228aSAlexey Bataev } else { 506aa7228aSAlexey Bataev std::cout << "Succeeded" << std::endl; 516aa7228aSAlexey Bataev } 526aa7228aSAlexey Bataev 536aa7228aSAlexey Bataev return 0; 546aa7228aSAlexey Bataev } 55