15ffd83dbSDimitry Andric /*===---- openmp_wrapper/math.h -------- OpenMP math.h intercept ------ c++ -=== 20b57cec5SDimitry Andric * 30b57cec5SDimitry Andric * Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. 40b57cec5SDimitry Andric * See https://llvm.org/LICENSE.txt for license information. 50b57cec5SDimitry Andric * SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception 60b57cec5SDimitry Andric * 70b57cec5SDimitry Andric *===-----------------------------------------------------------------------=== 80b57cec5SDimitry Andric */ 90b57cec5SDimitry Andric 105ffd83dbSDimitry Andric // If we are in C++ mode and include <math.h> (not <cmath>) first, we still need 115ffd83dbSDimitry Andric // to make sure <cmath> is read first. The problem otherwise is that we haven't 125ffd83dbSDimitry Andric // seen the declarations of the math.h functions when the system math.h includes 135ffd83dbSDimitry Andric // our cmath overlay. However, our cmath overlay, or better the underlying 145ffd83dbSDimitry Andric // overlay, e.g. CUDA, uses the math.h functions. Since we haven't declared them 155ffd83dbSDimitry Andric // yet we get errors. CUDA avoids this by eagerly declaring all math functions 165ffd83dbSDimitry Andric // (in the __device__ space) but we cannot do this. Instead we break the 175ffd83dbSDimitry Andric // dependence by forcing cmath to go first. While our cmath will in turn include 185ffd83dbSDimitry Andric // this file, the cmath guards will prevent recursion. 195ffd83dbSDimitry Andric #ifdef __cplusplus 205ffd83dbSDimitry Andric #include <cmath> 210b57cec5SDimitry Andric #endif 220b57cec5SDimitry Andric 235ffd83dbSDimitry Andric #ifndef __CLANG_OPENMP_MATH_H__ 245ffd83dbSDimitry Andric #define __CLANG_OPENMP_MATH_H__ 255ffd83dbSDimitry Andric 265ffd83dbSDimitry Andric #ifndef _OPENMP 275ffd83dbSDimitry Andric #error "This file is for OpenMP compilation only." 285ffd83dbSDimitry Andric #endif 295ffd83dbSDimitry Andric 305ffd83dbSDimitry Andric #include_next <math.h> 315ffd83dbSDimitry Andric 325ffd83dbSDimitry Andric // We need limits.h for __clang_cuda_math.h below and because it should not hurt 335ffd83dbSDimitry Andric // we include it eagerly here. 345ffd83dbSDimitry Andric #include <limits.h> 355ffd83dbSDimitry Andric 365ffd83dbSDimitry Andric // We need stdlib.h because (for now) __clang_cuda_math.h below declares `abs` 375ffd83dbSDimitry Andric // which should live in stdlib.h. 385ffd83dbSDimitry Andric #include <stdlib.h> 395ffd83dbSDimitry Andric 405ffd83dbSDimitry Andric #pragma omp begin declare variant match( \ 415ffd83dbSDimitry Andric device = {arch(nvptx, nvptx64)}, implementation = {extension(match_any)}) 425ffd83dbSDimitry Andric 435ffd83dbSDimitry Andric #define __CUDA__ 445ffd83dbSDimitry Andric #define __OPENMP_NVPTX__ 455ffd83dbSDimitry Andric #include <__clang_cuda_math.h> 465ffd83dbSDimitry Andric #undef __OPENMP_NVPTX__ 475ffd83dbSDimitry Andric #undef __CUDA__ 485ffd83dbSDimitry Andric 495ffd83dbSDimitry Andric #pragma omp end declare variant 505ffd83dbSDimitry Andric 51*69ade1e0SDimitry Andric #ifdef __AMDGCN__ 52*69ade1e0SDimitry Andric #pragma omp begin declare variant match(device = {arch(amdgcn)}) 53*69ade1e0SDimitry Andric 54*69ade1e0SDimitry Andric #define __OPENMP_AMDGCN__ 55*69ade1e0SDimitry Andric #include <__clang_hip_math.h> 56*69ade1e0SDimitry Andric #undef __OPENMP_AMDGCN__ 57*69ade1e0SDimitry Andric 58*69ade1e0SDimitry Andric #pragma omp end declare variant 59*69ade1e0SDimitry Andric #endif 60*69ade1e0SDimitry Andric 615ffd83dbSDimitry Andric #endif 62