144 lines · c
1// clang-format off2// RUN: %libomptarget-compile-run-and-check-generic3// REQUIRES: ompt4// REQUIRES: gpu5// clang-format on6 7/*8 * Example OpenMP program that registers EMI callbacks9 */10 11#include <assert.h>12#include <omp.h>13#include <stdio.h>14 15#include "callbacks.h"16#include "register_emi.h"17 18int main() {19 int N = 100000;20 21 int a[N];22 int b[N];23 24 int i;25 26 for (i = 0; i < N; i++)27 a[i] = 0;28 29 for (i = 0; i < N; i++)30 b[i] = i;31 32#pragma omp target parallel for33 {34 for (int j = 0; j < N; j++)35 a[j] = b[j];36 }37 38#pragma omp target teams distribute parallel for39 {40 for (int j = 0; j < N; j++)41 a[j] = b[j];42 }43 44 int rc = 0;45 for (i = 0; i < N; i++)46 if (a[i] != b[i]) {47 rc++;48 printf("Wrong value: a[%d]=%d\n", i, a[i]);49 }50 51 if (!rc)52 printf("Success\n");53 54 return rc;55}56 57// clang-format off58/// CHECK: Callback Init:59/// CHECK: Callback Load:60/// CHECK: Callback Target EMI: kind=ompt_target endpoint=ompt_scope_begin61/// CHECK-NOT: code=(nil)62/// CHECK: code=[[CODE1:.*]]63/// CHECK: Callback DataOp EMI: endpoint=ompt_scope_begin optype=ompt_target_data_alloc64/// CHECK: code=[[CODE1]]65/// CHECK: Callback DataOp EMI: endpoint=ompt_scope_end optype=ompt_target_data_alloc66/// CHECK-NOT: dest=(nil)67/// CHECK: code=[[CODE1]]68/// CHECK: Callback DataOp EMI: endpoint=ompt_scope_begin optype=ompt_target_data_transfer_to_device69/// CHECK: code=[[CODE1]]70/// CHECK: Callback DataOp EMI: endpoint=ompt_scope_end optype=ompt_target_data_transfer_to_device71/// CHECK: code=[[CODE1]]72/// CHECK: Callback DataOp EMI: endpoint=ompt_scope_begin optype=ompt_target_data_alloc73/// CHECK: code=[[CODE1]]74/// CHECK: Callback DataOp EMI: endpoint=ompt_scope_end optype=ompt_target_data_alloc75/// CHECK-NOT: dest=(nil)76/// CHECK: code=[[CODE1]]77/// CHECK: Callback DataOp EMI: endpoint=ompt_scope_begin optype=ompt_target_data_transfer_to_device78/// CHECK: code=[[CODE1]]79/// CHECK: Callback DataOp EMI: endpoint=ompt_scope_end optype=ompt_target_data_transfer_to_device80/// CHECK: code=[[CODE1]]81/// CHECK: Callback Submit EMI: endpoint=ompt_scope_begin req_num_teams=182/// CHECK: Callback Submit EMI: endpoint=ompt_scope_end req_num_teams=183/// CHECK: Callback DataOp EMI: endpoint=ompt_scope_begin optype=ompt_target_data_transfer_from_device84/// CHECK: code=[[CODE1]]85/// CHECK: Callback DataOp EMI: endpoint=ompt_scope_end optype=ompt_target_data_transfer_from_device86/// CHECK: code=[[CODE1]]87/// CHECK: Callback DataOp EMI: endpoint=ompt_scope_begin optype=ompt_target_data_transfer_from_device88/// CHECK: code=[[CODE1]]89/// CHECK: Callback DataOp EMI: endpoint=ompt_scope_end optype=ompt_target_data_transfer_from_device90/// CHECK: code=[[CODE1]]91/// CHECK: Callback DataOp EMI: endpoint=ompt_scope_begin optype=ompt_target_data_delete92/// CHECK: code=[[CODE1]]93/// CHECK: Callback DataOp EMI: endpoint=ompt_scope_end optype=ompt_target_data_delete94/// CHECK: code=[[CODE1]]95/// CHECK: Callback DataOp EMI: endpoint=ompt_scope_begin optype=ompt_target_data_delete96/// CHECK: code=[[CODE1]]97/// CHECK: Callback DataOp EMI: endpoint=ompt_scope_end optype=ompt_target_data_delete98/// CHECK: code=[[CODE1]]99/// CHECK: Callback Target EMI: kind=ompt_target endpoint=ompt_scope_end100/// CHECK: code=[[CODE1]]101 102/// CHECK: Callback Target EMI: kind=ompt_target endpoint=ompt_scope_begin103/// CHECK-NOT: code=(nil)104/// CHECK: code=[[CODE2:.*]]105/// CHECK: Callback DataOp EMI: endpoint=ompt_scope_begin optype=ompt_target_data_alloc106/// CHECK: code=[[CODE2]]107/// CHECK: Callback DataOp EMI: endpoint=ompt_scope_end optype=ompt_target_data_alloc108/// CHECK-NOT: dest=(nil)109/// CHECK: code=[[CODE2]]110/// CHECK: Callback DataOp EMI: endpoint=ompt_scope_begin optype=ompt_target_data_transfer_to_device111/// CHECK: code=[[CODE2]]112/// CHECK: Callback DataOp EMI: endpoint=ompt_scope_end optype=ompt_target_data_transfer_to_device113/// CHECK: code=[[CODE2]]114/// CHECK: Callback DataOp EMI: endpoint=ompt_scope_begin optype=ompt_target_data_alloc115/// CHECK: code=[[CODE2]]116/// CHECK: Callback DataOp EMI: endpoint=ompt_scope_end optype=ompt_target_data_alloc117/// CHECK-NOT: dest=(nil)118/// CHECK: code=[[CODE2]]119/// CHECK: Callback DataOp EMI: endpoint=ompt_scope_begin optype=ompt_target_data_transfer_to_device120/// CHECK: code=[[CODE2]]121/// CHECK: Callback DataOp EMI: endpoint=ompt_scope_end optype=ompt_target_data_transfer_to_device122/// CHECK: code=[[CODE2]]123/// CHECK: Callback Submit EMI: endpoint=ompt_scope_begin req_num_teams=0124/// CHECK: Callback Submit EMI: endpoint=ompt_scope_end req_num_teams=0125/// CHECK: Callback DataOp EMI: endpoint=ompt_scope_begin optype=ompt_target_data_transfer_from_device126/// CHECK: code=[[CODE2]]127/// CHECK: Callback DataOp EMI: endpoint=ompt_scope_end optype=ompt_target_data_transfer_from_device128/// CHECK: code=[[CODE2]]129/// CHECK: Callback DataOp EMI: endpoint=ompt_scope_begin optype=ompt_target_data_transfer_from_device130/// CHECK: code=[[CODE2]]131/// CHECK: Callback DataOp EMI: endpoint=ompt_scope_end optype=ompt_target_data_transfer_from_device132/// CHECK: code=[[CODE2]]133/// CHECK: Callback DataOp EMI: endpoint=ompt_scope_begin optype=ompt_target_data_delete134/// CHECK: code=[[CODE2]]135/// CHECK: Callback DataOp EMI: endpoint=ompt_scope_end optype=ompt_target_data_delete136/// CHECK: code=[[CODE2]]137/// CHECK: Callback DataOp EMI: endpoint=ompt_scope_begin optype=ompt_target_data_delete138/// CHECK: code=[[CODE2]]139/// CHECK: Callback DataOp EMI: endpoint=ompt_scope_end optype=ompt_target_data_delete140/// CHECK: code=[[CODE2]]141/// CHECK: Callback Target EMI: kind=ompt_target endpoint=ompt_scope_end142/// CHECK: code=[[CODE2]]143/// CHECK: Callback Fini:144