Compare commits
1
Commits
master
..
release_90
| Author | SHA1 | Date | |
|---|---|---|---|
 Hans Wennborg
|
b76842ed16 |
+9
-6
@@ -133,7 +133,7 @@ Options for all Libraries
|
||||
Options for ``libomp``
|
||||
----------------------
|
||||
|
||||
**LIBOMP_ARCH** = ``aarch64|arm|i386|mic|mips|mips64|ppc64|ppc64le|x86_64|riscv64``
|
||||
**LIBOMP_ARCH** = ``aarch64|arm|i386|mic|mips|mips64|ppc64|ppc64le|x86_64``
|
||||
The default value for this option is chosen based on probing the compiler for
|
||||
architecture macros (e.g., is ``__x86_64__`` predefined by compiler?).
|
||||
|
||||
@@ -189,8 +189,8 @@ Optional Features
|
||||
|
||||
**LIBOMP_OMPT_SUPPORT** = ``ON|OFF``
|
||||
Include support for the OpenMP Tools Interface (OMPT).
|
||||
This option is supported and ``ON`` by default for x86, x86_64, AArch64,
|
||||
PPC64 and RISCV64 on Linux* and macOS*.
|
||||
This option is supported and ``ON`` by default for x86, x86_64, AArch64, and
|
||||
PPC64 on Linux* and macOS*.
|
||||
This option is ``OFF`` if this feature is not supported for the platform.
|
||||
|
||||
**LIBOMP_OMPT_OPTIONAL** = ``ON|OFF``
|
||||
@@ -221,6 +221,9 @@ These flags are **appended**, they do not overwrite any of the preset flags.
|
||||
**LIBOMP_CPPFLAGS** = <space-separated flags>
|
||||
Additional C preprocessor flags.
|
||||
|
||||
**LIBOMP_CFLAGS** = <space-separated flags>
|
||||
Additional C compiler flags.
|
||||
|
||||
**LIBOMP_CXXFLAGS** = <space-separated flags>
|
||||
Additional C++ compiler flags.
|
||||
|
||||
@@ -318,12 +321,12 @@ Advanced Builds with Various Options
|
||||
|
||||
$ cmake -DCMAKE_C_COMPILER=icc -DCMAKE_CXX_COMPILER=icpc -DCMAKE_Fortran_COMPILER=ifort -DLIBOMP_FORTRAN_MODULES=on ..
|
||||
|
||||
- Have CMake find the C/C++ compiler and specify additional flags for the
|
||||
preprocessor and C++ compiler.
|
||||
- Have CMake find the C/C++ compiler and specify additional flags for the C
|
||||
compiler, preprocessor, and C++ compiler.
|
||||
|
||||
.. code-blocks:: console
|
||||
|
||||
$ cmake -DLIBOMP_CPPFLAGS='-DNEW_FEATURE=1 -DOLD_FEATURE=0' -DLIBOMP_CXXFLAGS='--one-specific-flag --two-specific-flag' ..
|
||||
$ cmake -DLIBOMP_CFLAGS='-specific-flag' -DLIBOMP_CPPFLAGS='-DNEW_FEATURE=1 -DOLD_FEATURE=0' -DLIBOMP_CXXFLAGS='--one-specific-flag --two-specific-flag' ..
|
||||
|
||||
- Build the stubs library
|
||||
|
||||
|
||||
@@ -1,4 +1,4 @@
|
||||
if (OPENMP_STANDALONE_BUILD)
|
||||
if (${OPENMP_STANDALONE_BUILD})
|
||||
# From HandleLLVMOptions.cmake
|
||||
function(append_if condition value)
|
||||
if (${condition})
|
||||
@@ -9,26 +9,10 @@ if (OPENMP_STANDALONE_BUILD)
|
||||
endfunction()
|
||||
endif()
|
||||
|
||||
# MSVC and clang-cl in compatibility mode map -Wall to -Weverything.
|
||||
# TODO: LLVM adds /W4 instead, check if that works for the OpenMP runtimes.
|
||||
if (NOT MSVC)
|
||||
append_if(OPENMP_HAVE_WALL_FLAG "-Wall" CMAKE_C_FLAGS CMAKE_CXX_FLAGS)
|
||||
endif()
|
||||
if (OPENMP_ENABLE_WERROR)
|
||||
if (${OPENMP_ENABLE_WERROR})
|
||||
append_if(OPENMP_HAVE_WERROR_FLAG "-Werror" CMAKE_C_FLAGS CMAKE_CXX_FLAGS)
|
||||
endif()
|
||||
|
||||
# Additional warnings that are not enabled by -Wall.
|
||||
append_if(OPENMP_HAVE_WCAST_QUAL_FLAG "-Wcast-qual" CMAKE_C_FLAGS CMAKE_CXX_FLAGS)
|
||||
append_if(OPENMP_HAVE_WFORMAT_PEDANTIC_FLAG "-Wformat-pedantic" CMAKE_C_FLAGS CMAKE_CXX_FLAGS)
|
||||
append_if(OPENMP_HAVE_WIMPLICIT_FALLTHROUGH_FLAG "-Wimplicit-fallthrough" CMAKE_C_FLAGS CMAKE_CXX_FLAGS)
|
||||
append_if(OPENMP_HAVE_WSIGN_COMPARE_FLAG "-Wsign-compare" CMAKE_C_FLAGS CMAKE_CXX_FLAGS)
|
||||
|
||||
# Warnings that we want to disable because they are too verbose or fragile.
|
||||
append_if(OPENMP_HAVE_WNO_EXTRA_FLAG "-Wno-extra" CMAKE_C_FLAGS CMAKE_CXX_FLAGS)
|
||||
append_if(OPENMP_HAVE_WNO_PEDANTIC_FLAG "-Wno-pedantic" CMAKE_C_FLAGS CMAKE_CXX_FLAGS)
|
||||
append_if(OPENMP_HAVE_WNO_MAYBE_UNINITIALIZED_FLAG "-Wno-maybe-uninitialized" CMAKE_C_FLAGS CMAKE_CXX_FLAGS)
|
||||
|
||||
append_if(OPENMP_HAVE_STD_GNUPP11_FLAG "-std=gnu++11" CMAKE_CXX_FLAGS)
|
||||
if (NOT OPENMP_HAVE_STD_GNUPP11_FLAG)
|
||||
append_if(OPENMP_HAVE_STD_CPP11_FLAG "-std=c++11" CMAKE_CXX_FLAGS)
|
||||
|
||||
+2
-13
@@ -1,18 +1,7 @@
|
||||
include(CheckCCompilerFlag)
|
||||
include(CheckCXXCompilerFlag)
|
||||
|
||||
check_cxx_compiler_flag(-Wall OPENMP_HAVE_WALL_FLAG)
|
||||
check_cxx_compiler_flag(-Werror OPENMP_HAVE_WERROR_FLAG)
|
||||
|
||||
# Additional warnings that are not enabled by -Wall.
|
||||
check_cxx_compiler_flag(-Wcast-qual OPENMP_HAVE_WCAST_QUAL_FLAG)
|
||||
check_cxx_compiler_flag(-Wformat-pedantic OPENMP_HAVE_WFORMAT_PEDANTIC_FLAG)
|
||||
check_cxx_compiler_flag(-Wimplicit-fallthrough OPENMP_HAVE_WIMPLICIT_FALLTHROUGH_FLAG)
|
||||
check_cxx_compiler_flag(-Wsign-compare OPENMP_HAVE_WSIGN_COMPARE_FLAG)
|
||||
|
||||
# Warnings that we want to disable because they are too verbose or fragile.
|
||||
check_cxx_compiler_flag(-Wno-extra OPENMP_HAVE_WNO_EXTRA_FLAG)
|
||||
check_cxx_compiler_flag(-Wno-pedantic OPENMP_HAVE_WNO_PEDANTIC_FLAG)
|
||||
check_cxx_compiler_flag(-Wno-maybe-uninitialized OPENMP_HAVE_WNO_MAYBE_UNINITIALIZED_FLAG)
|
||||
check_c_compiler_flag(-Werror OPENMP_HAVE_WERROR_FLAG)
|
||||
|
||||
check_cxx_compiler_flag(-std=gnu++11 OPENMP_HAVE_STD_GNUPP11_FLAG)
|
||||
check_cxx_compiler_flag(-std=c++11 OPENMP_HAVE_STD_CPP11_FLAG)
|
||||
|
||||
@@ -78,7 +78,7 @@ endfunction()
|
||||
|
||||
# These flags are required to emit LLVM Bitcode. We check them together because
|
||||
# if any of them are not supported, there is no point in finding out which are.
|
||||
set(compiler_flags_required -emit-llvm -O1 --cuda-device-only -std=c++11 --cuda-path=${CUDA_TOOLKIT_ROOT_DIR})
|
||||
set(compiler_flags_required -emit-llvm -O1 --cuda-device-only --cuda-path=${CUDA_TOOLKIT_ROOT_DIR})
|
||||
set(compiler_flags_required_src "extern \"C\" __device__ int thread() { return threadIdx.x; }")
|
||||
check_bitcode_compilation(LIBOMPTARGET_NVPTX_CUDA_COMPILER_SUPPORTS_FLAGS_REQUIRED "${compiler_flags_required_src}" ${compiler_flags_required})
|
||||
|
||||
|
||||
@@ -35,10 +35,6 @@ if(CUDA_HOST_COMPILER MATCHES clang)
|
||||
set(CUDA_HOST_COMPILER "${LIBOMPTARGET_NVPTX_ALTERNATE_GCC_HOST_COMPILER}" CACHE FILEPATH "" FORCE)
|
||||
endif()
|
||||
|
||||
get_filename_component(devicertl_base_directory
|
||||
${CMAKE_CURRENT_SOURCE_DIR}
|
||||
DIRECTORY)
|
||||
|
||||
if(LIBOMPTARGET_DEP_CUDA_FOUND)
|
||||
libomptarget_say("Building CUDA offloading device RTL.")
|
||||
|
||||
@@ -87,7 +83,7 @@ if(LIBOMPTARGET_DEP_CUDA_FOUND)
|
||||
# yet supported by the CUDA toolchain on the device.
|
||||
set(BUILD_SHARED_LIBS OFF)
|
||||
set(CUDA_SEPARABLE_COMPILATION ON)
|
||||
list(APPEND CUDA_NVCC_FLAGS -I${devicertl_base_directory})
|
||||
|
||||
cuda_add_library(omptarget-nvptx STATIC ${cuda_src_files} ${omp_data_objects}
|
||||
OPTIONS ${CUDA_ARCH} ${CUDA_DEBUG})
|
||||
|
||||
@@ -121,8 +117,7 @@ if(LIBOMPTARGET_DEP_CUDA_FOUND)
|
||||
libomptarget_say("Building CUDA LLVM bitcode offloading device RTL.")
|
||||
|
||||
# Set flags for LLVM Bitcode compilation.
|
||||
set(bc_flags ${LIBOMPTARGET_NVPTX_SELECTED_CUDA_COMPILER_FLAGS}
|
||||
-I${devicertl_base_directory})
|
||||
set(bc_flags ${LIBOMPTARGET_NVPTX_SELECTED_CUDA_COMPILER_FLAGS})
|
||||
if(${LIBOMPTARGET_NVPTX_DEBUG})
|
||||
set(bc_flags ${bc_flags} -DOMPTARGET_NVPTX_DEBUG=-1)
|
||||
else()
|
||||
|
||||
@@ -10,7 +10,6 @@
|
||||
//
|
||||
//===----------------------------------------------------------------------===//
|
||||
#include "omptarget-nvptx.h"
|
||||
#include "target_impl.h"
|
||||
#include <stdio.h>
|
||||
|
||||
// Warp ID in the CUDA block
|
||||
@@ -20,7 +19,7 @@ INLINE static unsigned getLaneId() { return threadIdx.x % WARPSIZE; }
|
||||
|
||||
// Return true if this is the first active thread in the warp.
|
||||
INLINE static bool IsWarpMasterActiveThread() {
|
||||
unsigned long long Mask = __kmpc_impl_activemask();
|
||||
unsigned long long Mask = __ACTIVEMASK();
|
||||
unsigned long long ShNum = WARPSIZE - (GetThreadIdInBlock() % WARPSIZE);
|
||||
unsigned long long Sh = Mask << ShNum;
|
||||
// Truncate Sh to the 32 lower bits
|
||||
@@ -96,7 +95,7 @@ __kmpc_initialize_data_sharing_environment(__kmpc_data_sharing_slot *rootS,
|
||||
|
||||
EXTERN void *__kmpc_data_sharing_environment_begin(
|
||||
__kmpc_data_sharing_slot **SavedSharedSlot, void **SavedSharedStack,
|
||||
void **SavedSharedFrame, __kmpc_impl_lanemask_t *SavedActiveThreads,
|
||||
void **SavedSharedFrame, int32_t *SavedActiveThreads,
|
||||
size_t SharingDataSize, size_t SharingDefaultDataSize,
|
||||
int16_t IsOMPRuntimeInitialized) {
|
||||
|
||||
@@ -112,12 +111,12 @@ EXTERN void *__kmpc_data_sharing_environment_begin(
|
||||
(unsigned long long)SharingDefaultDataSize);
|
||||
|
||||
unsigned WID = getWarpId();
|
||||
__kmpc_impl_lanemask_t CurActiveThreads = __kmpc_impl_activemask();
|
||||
unsigned CurActiveThreads = __ACTIVEMASK();
|
||||
|
||||
__kmpc_data_sharing_slot *&SlotP = DataSharingState.SlotPtr[WID];
|
||||
void *&StackP = DataSharingState.StackPtr[WID];
|
||||
void * volatile &FrameP = DataSharingState.FramePtr[WID];
|
||||
__kmpc_impl_lanemask_t &ActiveT = DataSharingState.ActiveThreads[WID];
|
||||
int32_t &ActiveT = DataSharingState.ActiveThreads[WID];
|
||||
|
||||
DSPRINT0(DSFLAG, "Save current slot/stack values.\n");
|
||||
// Save the current values.
|
||||
@@ -225,7 +224,7 @@ EXTERN void *__kmpc_data_sharing_environment_begin(
|
||||
|
||||
EXTERN void __kmpc_data_sharing_environment_end(
|
||||
__kmpc_data_sharing_slot **SavedSharedSlot, void **SavedSharedStack,
|
||||
void **SavedSharedFrame, __kmpc_impl_lanemask_t *SavedActiveThreads,
|
||||
void **SavedSharedFrame, int32_t *SavedActiveThreads,
|
||||
int32_t IsEntryPoint) {
|
||||
|
||||
DSPRINT0(DSFLAG, "Entering __kmpc_data_sharing_environment_end\n");
|
||||
@@ -252,7 +251,7 @@ EXTERN void __kmpc_data_sharing_environment_end(
|
||||
return;
|
||||
}
|
||||
|
||||
__kmpc_impl_lanemask_t CurActive = __kmpc_impl_activemask();
|
||||
int32_t CurActive = __ACTIVEMASK();
|
||||
|
||||
// Only the warp master can restore the stack and frame information, and only
|
||||
// if there are no other threads left behind in this environment (i.e. the
|
||||
@@ -260,7 +259,7 @@ EXTERN void __kmpc_data_sharing_environment_end(
|
||||
// assume that threads will converge right after the call site that started
|
||||
// the environment.
|
||||
if (IsWarpMasterActiveThread()) {
|
||||
__kmpc_impl_lanemask_t &ActiveT = DataSharingState.ActiveThreads[WID];
|
||||
int32_t &ActiveT = DataSharingState.ActiveThreads[WID];
|
||||
|
||||
DSPRINT0(DSFLAG, "Before restoring the stack\n");
|
||||
// Zero the bits in the mask. If it is still different from zero, then we
|
||||
@@ -378,7 +377,7 @@ INLINE static void* data_sharing_push_stack_common(size_t PushSize) {
|
||||
// Frame pointer must be visible to all workers in the same warp.
|
||||
const unsigned WID = getWarpId();
|
||||
void *FrameP = 0;
|
||||
__kmpc_impl_lanemask_t CurActive = __kmpc_impl_activemask();
|
||||
int32_t CurActive = __ACTIVEMASK();
|
||||
|
||||
if (IsWarpMaster) {
|
||||
// SlotP will point to either the shared memory slot or an existing
|
||||
@@ -431,10 +430,9 @@ INLINE static void* data_sharing_push_stack_common(size_t PushSize) {
|
||||
}
|
||||
}
|
||||
// Get address from lane 0.
|
||||
int *FP = (int *)&FrameP;
|
||||
FP[0] = __kmpc_impl_shfl_sync(CurActive, FP[0], 0);
|
||||
((int *)&FrameP)[0] = __SHFL_SYNC(CurActive, ((int *)&FrameP)[0], 0);
|
||||
if (sizeof(FrameP) == 8)
|
||||
FP[1] = __kmpc_impl_shfl_sync(CurActive, FP[1], 0);
|
||||
((int *)&FrameP)[1] = __SHFL_SYNC(CurActive, ((int *)&FrameP)[1], 0);
|
||||
|
||||
return FrameP;
|
||||
}
|
||||
@@ -553,7 +551,8 @@ EXTERN void __kmpc_get_team_static_memory(int16_t isSPMDExecutionMode,
|
||||
if (GetThreadIdInBlock() == 0) {
|
||||
*frame = omptarget_nvptx_simpleMemoryManager.Acquire(buf, size);
|
||||
}
|
||||
__kmpc_impl_syncthreads();
|
||||
// FIXME: use __syncthreads instead when the function copy is fixed in LLVM.
|
||||
__SYNCTHREADS();
|
||||
return;
|
||||
}
|
||||
ASSERT0(LT_FUSSY, GetThreadIdInBlock() == GetMasterThreadID(),
|
||||
@@ -567,7 +566,8 @@ EXTERN void __kmpc_restore_team_static_memory(int16_t isSPMDExecutionMode,
|
||||
if (is_shared)
|
||||
return;
|
||||
if (isSPMDExecutionMode) {
|
||||
__kmpc_impl_syncthreads();
|
||||
// FIXME: use __syncthreads instead when the function copy is fixed in LLVM.
|
||||
__SYNCTHREADS();
|
||||
if (GetThreadIdInBlock() == 0) {
|
||||
omptarget_nvptx_simpleMemoryManager.Release();
|
||||
}
|
||||
|
||||
+9
-16
@@ -1,4 +1,4 @@
|
||||
//===------- interface.h - OpenMP interface definitions ---------- CUDA -*-===//
|
||||
//===------- interface.h - NVPTX OpenMP interface definitions ---- CUDA -*-===//
|
||||
//
|
||||
// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
|
||||
// See https://llvm.org/LICENSE.txt for license information.
|
||||
@@ -6,6 +6,8 @@
|
||||
//
|
||||
//===----------------------------------------------------------------------===//
|
||||
//
|
||||
// This file contains debug macros to be used in the application.
|
||||
//
|
||||
// This file contains all the definitions that are relevant to
|
||||
// the interface. The first section contains the interface as
|
||||
// declared by OpenMP. The second section includes the compiler
|
||||
@@ -16,11 +18,7 @@
|
||||
#ifndef _INTERFACES_H_
|
||||
#define _INTERFACES_H_
|
||||
|
||||
#include <stdint.h>
|
||||
|
||||
#ifdef __CUDACC__
|
||||
#include "nvptx/src/nvptx_interface.h"
|
||||
#endif
|
||||
#include "option.h"
|
||||
|
||||
////////////////////////////////////////////////////////////////////////////////
|
||||
// OpenMP interface
|
||||
@@ -424,9 +422,7 @@ EXTERN void __kmpc_end_critical(kmp_Ident *loc, int32_t global_tid,
|
||||
EXTERN void __kmpc_flush(kmp_Ident *loc);
|
||||
|
||||
// vote
|
||||
EXTERN __kmpc_impl_lanemask_t __kmpc_warp_active_thread_mask();
|
||||
// syncwarp
|
||||
EXTERN void __kmpc_syncwarp(__kmpc_impl_lanemask_t);
|
||||
EXTERN int32_t __kmpc_warp_active_thread_mask();
|
||||
|
||||
// tasks
|
||||
EXTERN kmp_TaskDescr *__kmpc_omp_task_alloc(kmp_Ident *loc,
|
||||
@@ -477,13 +473,11 @@ EXTERN void __kmpc_kernel_prepare_parallel(void *WorkFn,
|
||||
EXTERN bool __kmpc_kernel_parallel(void **WorkFn,
|
||||
int16_t IsOMPRuntimeInitialized);
|
||||
EXTERN void __kmpc_kernel_end_parallel();
|
||||
EXTERN bool __kmpc_kernel_convergent_parallel(void *buffer,
|
||||
__kmpc_impl_lanemask_t Mask,
|
||||
EXTERN bool __kmpc_kernel_convergent_parallel(void *buffer, uint32_t Mask,
|
||||
bool *IsFinal,
|
||||
int32_t *LaneSource);
|
||||
EXTERN void __kmpc_kernel_end_convergent_parallel(void *buffer);
|
||||
EXTERN bool __kmpc_kernel_convergent_simd(void *buffer,
|
||||
__kmpc_impl_lanemask_t Mask,
|
||||
EXTERN bool __kmpc_kernel_convergent_simd(void *buffer, uint32_t Mask,
|
||||
bool *IsFinal, int32_t *LaneSource,
|
||||
int32_t *LaneId, int32_t *NumLanes);
|
||||
EXTERN void __kmpc_kernel_end_convergent_simd(void *buffer);
|
||||
@@ -514,13 +508,12 @@ __kmpc_initialize_data_sharing_environment(__kmpc_data_sharing_slot *RootS,
|
||||
size_t InitialDataSize);
|
||||
EXTERN void *__kmpc_data_sharing_environment_begin(
|
||||
__kmpc_data_sharing_slot **SavedSharedSlot, void **SavedSharedStack,
|
||||
void **SavedSharedFrame, __kmpc_impl_lanemask_t *SavedActiveThreads,
|
||||
void **SavedSharedFrame, int32_t *SavedActiveThreads,
|
||||
size_t SharingDataSize, size_t SharingDefaultDataSize,
|
||||
int16_t IsOMPRuntimeInitialized);
|
||||
EXTERN void __kmpc_data_sharing_environment_end(
|
||||
__kmpc_data_sharing_slot **SavedSharedSlot, void **SavedSharedStack,
|
||||
void **SavedSharedFrame, __kmpc_impl_lanemask_t *SavedActiveThreads,
|
||||
int32_t IsEntryPoint);
|
||||
void **SavedSharedFrame, int32_t *SavedActiveThreads, int32_t IsEntryPoint);
|
||||
|
||||
EXTERN void *
|
||||
__kmpc_get_data_sharing_environment_frame(int32_t SourceThreadID,
|
||||
@@ -13,7 +13,6 @@
|
||||
//===----------------------------------------------------------------------===//
|
||||
|
||||
#include "omptarget-nvptx.h"
|
||||
#include "target_impl.h"
|
||||
|
||||
////////////////////////////////////////////////////////////////////////////////
|
||||
////////////////////////////////////////////////////////////////////////////////
|
||||
@@ -380,21 +379,22 @@ public:
|
||||
////////////////////////////////////////////////////////////////////////////////
|
||||
// Support for dispatch next
|
||||
|
||||
INLINE static uint64_t Shuffle(__kmpc_impl_lanemask_t active, int64_t val,
|
||||
int leader) {
|
||||
uint32_t lo, hi;
|
||||
__kmpc_impl_unpack(val, lo, hi);
|
||||
hi = __kmpc_impl_shfl_sync(active, hi, leader);
|
||||
lo = __kmpc_impl_shfl_sync(active, lo, leader);
|
||||
return __kmpc_impl_pack(lo, hi);
|
||||
INLINE static int64_t Shuffle(unsigned active, int64_t val, int leader) {
|
||||
int lo, hi;
|
||||
asm volatile("mov.b64 {%0,%1}, %2;" : "=r"(lo), "=r"(hi) : "l"(val));
|
||||
hi = __SHFL_SYNC(active, hi, leader);
|
||||
lo = __SHFL_SYNC(active, lo, leader);
|
||||
asm volatile("mov.b64 %0, {%1,%2};" : "=l"(val) : "r"(lo), "r"(hi));
|
||||
return val;
|
||||
}
|
||||
|
||||
INLINE static uint64_t NextIter() {
|
||||
__kmpc_impl_lanemask_t active = __kmpc_impl_activemask();
|
||||
uint32_t leader = __kmpc_impl_ffs(active) - 1;
|
||||
uint32_t change = __kmpc_impl_popc(active);
|
||||
__kmpc_impl_lanemask_t lane_mask_lt = __kmpc_impl_lanemask_lt();
|
||||
unsigned int rank = __kmpc_impl_popc(active & lane_mask_lt);
|
||||
unsigned int active = __ACTIVEMASK();
|
||||
int leader = __ffs(active) - 1;
|
||||
int change = __popc(active);
|
||||
unsigned lane_mask_lt;
|
||||
asm("mov.u32 %0, %%lanemask_lt;" : "=r"(lane_mask_lt));
|
||||
unsigned int rank = __popc(active & lane_mask_lt);
|
||||
uint64_t warp_res;
|
||||
if (rank == 0) {
|
||||
warp_res = atomicAdd(
|
||||
|
||||
@@ -1,17 +0,0 @@
|
||||
//===--- nvptx_interface.h - OpenMP interface definitions -------- CUDA -*-===//
|
||||
//
|
||||
// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
|
||||
// See https://llvm.org/LICENSE.txt for license information.
|
||||
// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
|
||||
//
|
||||
//===----------------------------------------------------------------------===//
|
||||
|
||||
#ifndef _NVPTX_INTERFACE_H_
|
||||
#define _NVPTX_INTERFACE_H_
|
||||
|
||||
#include <stdint.h>
|
||||
|
||||
#define EXTERN extern "C" __device__
|
||||
typedef uint32_t __kmpc_impl_lanemask_t;
|
||||
|
||||
#endif
|
||||
@@ -11,7 +11,6 @@
|
||||
//===----------------------------------------------------------------------===//
|
||||
|
||||
#include "omptarget-nvptx.h"
|
||||
#include "target_impl.h"
|
||||
|
||||
////////////////////////////////////////////////////////////////////////////////
|
||||
// global data tables
|
||||
@@ -107,7 +106,7 @@ EXTERN void __kmpc_spmd_kernel_init(int ThreadLimit, int16_t RequiresOMPRuntime,
|
||||
}
|
||||
if (!RequiresOMPRuntime) {
|
||||
// Runtime is not required - exit.
|
||||
__kmpc_impl_syncthreads();
|
||||
__SYNCTHREADS();
|
||||
return;
|
||||
}
|
||||
|
||||
@@ -126,7 +125,8 @@ EXTERN void __kmpc_spmd_kernel_init(int ThreadLimit, int16_t RequiresOMPRuntime,
|
||||
// init team context
|
||||
currTeamDescr.InitTeamDescr();
|
||||
}
|
||||
__kmpc_impl_syncthreads();
|
||||
// FIXME: use __syncthreads instead when the function copy is fixed in LLVM.
|
||||
__SYNCTHREADS();
|
||||
|
||||
omptarget_nvptx_TeamDescr &currTeamDescr = getMyTeamDescriptor();
|
||||
omptarget_nvptx_WorkDescr &workDescr = getMyWorkDescriptor();
|
||||
@@ -168,7 +168,8 @@ EXTERN void __kmpc_spmd_kernel_deinit_v2(int16_t RequiresOMPRuntime) {
|
||||
if (!RequiresOMPRuntime)
|
||||
return;
|
||||
|
||||
__kmpc_impl_syncthreads();
|
||||
// FIXME: use __syncthreads instead when the function copy is fixed in LLVM.
|
||||
__SYNCTHREADS();
|
||||
int threadId = GetThreadIdInBlock();
|
||||
if (threadId == 0) {
|
||||
// Enqueue omp state object for use by another team.
|
||||
|
||||
@@ -45,6 +45,31 @@
|
||||
#define BARRIER_COUNTER 0
|
||||
#define ORDERED_COUNTER 1
|
||||
|
||||
// Macros for Cuda intrinsics
|
||||
// In Cuda 9.0, the *_sync() version takes an extra argument 'mask'.
|
||||
// Also, __ballot(1) in Cuda 8.0 is replaced with __activemask().
|
||||
#ifndef CUDA_VERSION
|
||||
#error CUDA_VERSION macro is undefined, something wrong with cuda.
|
||||
#elif CUDA_VERSION >= 9000
|
||||
#define __SHFL_SYNC(mask, var, srcLane) __shfl_sync((mask), (var), (srcLane))
|
||||
#define __SHFL_DOWN_SYNC(mask, var, delta, width) \
|
||||
__shfl_down_sync((mask), (var), (delta), (width))
|
||||
#define __ACTIVEMASK() __activemask()
|
||||
#else
|
||||
#define __SHFL_SYNC(mask, var, srcLane) __shfl((var), (srcLane))
|
||||
#define __SHFL_DOWN_SYNC(mask, var, delta, width) \
|
||||
__shfl_down((var), (delta), (width))
|
||||
#define __ACTIVEMASK() __ballot(1)
|
||||
#endif // CUDA_VERSION
|
||||
|
||||
#define __SYNCTHREADS_N(n) asm volatile("bar.sync %0;" : : "r"(n) : "memory");
|
||||
// Use original __syncthreads if compiled by nvcc or clang >= 9.0.
|
||||
#if !defined(__clang__) || __clang_major__ >= 9
|
||||
#define __SYNCTHREADS() __syncthreads()
|
||||
#else
|
||||
#define __SYNCTHREADS() __SYNCTHREADS_N(0)
|
||||
#endif
|
||||
|
||||
// arguments needed for L0 parallelism only.
|
||||
class omptarget_nvptx_SharedArgs {
|
||||
public:
|
||||
@@ -107,7 +132,7 @@ struct DataSharingStateTy {
|
||||
__kmpc_data_sharing_slot *SlotPtr[DS_Max_Warp_Number];
|
||||
void *StackPtr[DS_Max_Warp_Number];
|
||||
void * volatile FramePtr[DS_Max_Warp_Number];
|
||||
__kmpc_impl_lanemask_t ActiveThreads[DS_Max_Warp_Number];
|
||||
int32_t ActiveThreads[DS_Max_Warp_Number];
|
||||
};
|
||||
// Additional worker slot type which is initialized with the default worker slot
|
||||
// size of 4*32 bytes.
|
||||
|
||||
@@ -12,8 +12,6 @@
|
||||
#ifndef _OPTION_H_
|
||||
#define _OPTION_H_
|
||||
|
||||
#include "interface.h"
|
||||
|
||||
////////////////////////////////////////////////////////////////////////////////
|
||||
// Kernel options
|
||||
////////////////////////////////////////////////////////////////////////////////
|
||||
@@ -56,7 +54,8 @@
|
||||
// misc options (by def everythig here is device)
|
||||
////////////////////////////////////////////////////////////////////////////////
|
||||
|
||||
#define INLINE __forceinline__ __device__
|
||||
#define EXTERN extern "C" __device__
|
||||
#define INLINE __inline__ __device__
|
||||
#define NOINLINE __noinline__ __device__
|
||||
#ifndef TRUE
|
||||
#define TRUE 1
|
||||
|
||||
@@ -33,7 +33,6 @@
|
||||
//===----------------------------------------------------------------------===//
|
||||
|
||||
#include "omptarget-nvptx.h"
|
||||
#include "target_impl.h"
|
||||
|
||||
typedef struct ConvergentSimdJob {
|
||||
omptarget_nvptx_TaskDescr taskDescr;
|
||||
@@ -44,18 +43,18 @@ typedef struct ConvergentSimdJob {
|
||||
////////////////////////////////////////////////////////////////////////////////
|
||||
// support for convergent simd (team of threads in a warp only)
|
||||
////////////////////////////////////////////////////////////////////////////////
|
||||
EXTERN bool __kmpc_kernel_convergent_simd(void *buffer,
|
||||
__kmpc_impl_lanemask_t Mask,
|
||||
EXTERN bool __kmpc_kernel_convergent_simd(void *buffer, uint32_t Mask,
|
||||
bool *IsFinal, int32_t *LaneSource,
|
||||
int32_t *LaneId, int32_t *NumLanes) {
|
||||
PRINT0(LD_IO, "call to __kmpc_kernel_convergent_simd\n");
|
||||
__kmpc_impl_lanemask_t ConvergentMask = Mask;
|
||||
int32_t ConvergentSize = __kmpc_impl_popc(ConvergentMask);
|
||||
__kmpc_impl_lanemask_t WorkRemaining = ConvergentMask >> (*LaneSource + 1);
|
||||
*LaneSource += __kmpc_impl_ffs(WorkRemaining);
|
||||
*IsFinal = __kmpc_impl_popc(WorkRemaining) == 1;
|
||||
__kmpc_impl_lanemask_t lanemask_lt = __kmpc_impl_lanemask_lt();
|
||||
*LaneId = __kmpc_impl_popc(ConvergentMask & lanemask_lt);
|
||||
uint32_t ConvergentMask = Mask;
|
||||
int32_t ConvergentSize = __popc(ConvergentMask);
|
||||
uint32_t WorkRemaining = ConvergentMask >> (*LaneSource + 1);
|
||||
*LaneSource += __ffs(WorkRemaining);
|
||||
*IsFinal = __popc(WorkRemaining) == 1;
|
||||
uint32_t lanemask_lt;
|
||||
asm("mov.u32 %0, %%lanemask_lt;" : "=r"(lanemask_lt));
|
||||
*LaneId = __popc(ConvergentMask & lanemask_lt);
|
||||
|
||||
int threadId = GetLogicalThreadIdInBlock(isSPMDMode());
|
||||
int sourceThreadId = (threadId & ~(WARPSIZE - 1)) + *LaneSource;
|
||||
@@ -65,7 +64,7 @@ EXTERN bool __kmpc_kernel_convergent_simd(void *buffer,
|
||||
omptarget_nvptx_threadPrivateContext->SimdLimitForNextSimd(threadId);
|
||||
job->slimForNextSimd = SimdLimit;
|
||||
|
||||
int32_t SimdLimitSource = __kmpc_impl_shfl_sync(Mask, SimdLimit, *LaneSource);
|
||||
int32_t SimdLimitSource = __SHFL_SYNC(Mask, SimdLimit, *LaneSource);
|
||||
// reset simdlimit to avoid propagating to successive #simd
|
||||
if (SimdLimitSource > 0 && threadId == sourceThreadId)
|
||||
omptarget_nvptx_threadPrivateContext->SimdLimitForNextSimd(threadId) = 0;
|
||||
@@ -118,18 +117,18 @@ typedef struct ConvergentParallelJob {
|
||||
////////////////////////////////////////////////////////////////////////////////
|
||||
// support for convergent parallelism (team of threads in a warp only)
|
||||
////////////////////////////////////////////////////////////////////////////////
|
||||
EXTERN bool __kmpc_kernel_convergent_parallel(void *buffer,
|
||||
__kmpc_impl_lanemask_t Mask,
|
||||
EXTERN bool __kmpc_kernel_convergent_parallel(void *buffer, uint32_t Mask,
|
||||
bool *IsFinal,
|
||||
int32_t *LaneSource) {
|
||||
PRINT0(LD_IO, "call to __kmpc_kernel_convergent_parallel\n");
|
||||
__kmpc_impl_lanemask_t ConvergentMask = Mask;
|
||||
int32_t ConvergentSize = __kmpc_impl_popc(ConvergentMask);
|
||||
__kmpc_impl_lanemask_t WorkRemaining = ConvergentMask >> (*LaneSource + 1);
|
||||
*LaneSource += __kmpc_impl_ffs(WorkRemaining);
|
||||
*IsFinal = __kmpc_impl_popc(WorkRemaining) == 1;
|
||||
__kmpc_impl_lanemask_t lanemask_lt = __kmpc_impl_lanemask_lt();
|
||||
uint32_t OmpId = __kmpc_impl_popc(ConvergentMask & lanemask_lt);
|
||||
uint32_t ConvergentMask = Mask;
|
||||
int32_t ConvergentSize = __popc(ConvergentMask);
|
||||
uint32_t WorkRemaining = ConvergentMask >> (*LaneSource + 1);
|
||||
*LaneSource += __ffs(WorkRemaining);
|
||||
*IsFinal = __popc(WorkRemaining) == 1;
|
||||
uint32_t lanemask_lt;
|
||||
asm("mov.u32 %0, %%lanemask_lt;" : "=r"(lanemask_lt));
|
||||
uint32_t OmpId = __popc(ConvergentMask & lanemask_lt);
|
||||
|
||||
int threadId = GetLogicalThreadIdInBlock(isSPMDMode());
|
||||
int sourceThreadId = (threadId & ~(WARPSIZE - 1)) + *LaneSource;
|
||||
@@ -139,8 +138,7 @@ EXTERN bool __kmpc_kernel_convergent_parallel(void *buffer,
|
||||
omptarget_nvptx_threadPrivateContext->NumThreadsForNextParallel(threadId);
|
||||
job->tnumForNextPar = NumThreadsClause;
|
||||
|
||||
int32_t NumThreadsSource =
|
||||
__kmpc_impl_shfl_sync(Mask, NumThreadsClause, *LaneSource);
|
||||
int32_t NumThreadsSource = __SHFL_SYNC(Mask, NumThreadsClause, *LaneSource);
|
||||
// reset numthreads to avoid propagating to successive #parallel
|
||||
if (NumThreadsSource > 0 && threadId == sourceThreadId)
|
||||
omptarget_nvptx_threadPrivateContext->NumThreadsForNextParallel(threadId) =
|
||||
@@ -313,16 +311,7 @@ EXTERN bool __kmpc_kernel_parallel(void **WorkFn,
|
||||
(int)newTaskDescr->ThreadId(), (int)nThreads);
|
||||
|
||||
isActive = true;
|
||||
// Reconverge the threads at the end of the parallel region to correctly
|
||||
// handle parallel levels.
|
||||
// In Cuda9+ in non-SPMD mode we have either 1 worker thread or the whole
|
||||
// warp. If only 1 thread is active, not need to reconverge the threads.
|
||||
// If we have the whole warp, reconverge all the threads in the warp before
|
||||
// actually trying to change the parallel level. Otherwise, parallel level
|
||||
// can be changed incorrectly because of threads divergence.
|
||||
bool IsActiveParallelRegion = threadsInTeam != 1;
|
||||
IncParallelLevel(IsActiveParallelRegion,
|
||||
IsActiveParallelRegion ? __kmpc_impl_all_lanes : 1u);
|
||||
IncParallelLevel(threadsInTeam != 1);
|
||||
}
|
||||
|
||||
return isActive;
|
||||
@@ -340,16 +329,7 @@ EXTERN void __kmpc_kernel_end_parallel() {
|
||||
omptarget_nvptx_threadPrivateContext->SetTopLevelTaskDescr(
|
||||
threadId, currTaskDescr->GetPrevTaskDescr());
|
||||
|
||||
// Reconverge the threads at the end of the parallel region to correctly
|
||||
// handle parallel levels.
|
||||
// In Cuda9+ in non-SPMD mode we have either 1 worker thread or the whole
|
||||
// warp. If only 1 thread is active, not need to reconverge the threads.
|
||||
// If we have the whole warp, reconverge all the threads in the warp before
|
||||
// actually trying to change the parallel level. Otherwise, parallel level can
|
||||
// be changed incorrectly because of threads divergence.
|
||||
bool IsActiveParallelRegion = threadsInTeam != 1;
|
||||
DecParallelLevel(IsActiveParallelRegion,
|
||||
IsActiveParallelRegion ? __kmpc_impl_all_lanes : 1u);
|
||||
DecParallelLevel(threadsInTeam != 1);
|
||||
}
|
||||
|
||||
////////////////////////////////////////////////////////////////////////////////
|
||||
@@ -359,7 +339,7 @@ EXTERN void __kmpc_kernel_end_parallel() {
|
||||
EXTERN void __kmpc_serialized_parallel(kmp_Ident *loc, uint32_t global_tid) {
|
||||
PRINT0(LD_IO, "call to __kmpc_serialized_parallel\n");
|
||||
|
||||
IncParallelLevel(/*ActiveParallel=*/false, __kmpc_impl_activemask());
|
||||
IncParallelLevel(/*ActiveParallel=*/false);
|
||||
|
||||
if (checkRuntimeUninitialized(loc)) {
|
||||
ASSERT0(LT_FUSSY, checkSPMDMode(loc),
|
||||
@@ -398,7 +378,7 @@ EXTERN void __kmpc_end_serialized_parallel(kmp_Ident *loc,
|
||||
uint32_t global_tid) {
|
||||
PRINT0(LD_IO, "call to __kmpc_end_serialized_parallel\n");
|
||||
|
||||
DecParallelLevel(/*ActiveParallel=*/false, __kmpc_impl_activemask());
|
||||
DecParallelLevel(/*ActiveParallel=*/false);
|
||||
|
||||
if (checkRuntimeUninitialized(loc)) {
|
||||
ASSERT0(LT_FUSSY, checkSPMDMode(loc),
|
||||
|
||||
@@ -15,7 +15,6 @@
|
||||
#include <stdio.h>
|
||||
|
||||
#include "omptarget-nvptx.h"
|
||||
#include "target_impl.h"
|
||||
|
||||
EXTERN
|
||||
void __kmpc_nvptx_end_reduce(int32_t global_tid) {}
|
||||
@@ -24,15 +23,16 @@ EXTERN
|
||||
void __kmpc_nvptx_end_reduce_nowait(int32_t global_tid) {}
|
||||
|
||||
EXTERN int32_t __kmpc_shuffle_int32(int32_t val, int16_t delta, int16_t size) {
|
||||
return __kmpc_impl_shfl_down_sync(__kmpc_impl_all_lanes, val, delta, size);
|
||||
return __SHFL_DOWN_SYNC(0xFFFFFFFF, val, delta, size);
|
||||
}
|
||||
|
||||
EXTERN int64_t __kmpc_shuffle_int64(int64_t val, int16_t delta, int16_t size) {
|
||||
uint32_t lo, hi;
|
||||
__kmpc_impl_unpack(val, lo, hi);
|
||||
hi = __kmpc_impl_shfl_down_sync(__kmpc_impl_all_lanes, hi, delta, size);
|
||||
lo = __kmpc_impl_shfl_down_sync(__kmpc_impl_all_lanes, lo, delta, size);
|
||||
return __kmpc_impl_pack(lo, hi);
|
||||
int lo, hi;
|
||||
asm volatile("mov.b64 {%0,%1}, %2;" : "=r"(lo), "=r"(hi) : "l"(val));
|
||||
hi = __SHFL_DOWN_SYNC(0xFFFFFFFF, hi, delta, size);
|
||||
lo = __SHFL_DOWN_SYNC(0xFFFFFFFF, lo, delta, size);
|
||||
asm volatile("mov.b64 %0, {%1,%2};" : "=l"(val) : "r"(lo), "r"(hi));
|
||||
return val;
|
||||
}
|
||||
|
||||
INLINE static void gpu_regular_warp_reduce(void *reduce_data,
|
||||
@@ -59,16 +59,18 @@ INLINE static void gpu_irregular_warp_reduce(void *reduce_data,
|
||||
|
||||
INLINE static uint32_t
|
||||
gpu_irregular_simd_reduce(void *reduce_data, kmp_ShuffleReductFctPtr shflFct) {
|
||||
uint32_t lanemask_lt;
|
||||
uint32_t lanemask_gt;
|
||||
uint32_t size, remote_id, physical_lane_id;
|
||||
physical_lane_id = GetThreadIdInBlock() % WARPSIZE;
|
||||
__kmpc_impl_lanemask_t lanemask_lt = __kmpc_impl_lanemask_lt();
|
||||
__kmpc_impl_lanemask_t Liveness = __kmpc_impl_activemask();
|
||||
uint32_t logical_lane_id = __kmpc_impl_popc(Liveness & lanemask_lt) * 2;
|
||||
__kmpc_impl_lanemask_t lanemask_gt = __kmpc_impl_lanemask_gt();
|
||||
asm("mov.u32 %0, %%lanemask_lt;" : "=r"(lanemask_lt));
|
||||
uint32_t Liveness = __ACTIVEMASK();
|
||||
uint32_t logical_lane_id = __popc(Liveness & lanemask_lt) * 2;
|
||||
asm("mov.u32 %0, %%lanemask_gt;" : "=r"(lanemask_gt));
|
||||
do {
|
||||
Liveness = __kmpc_impl_activemask();
|
||||
remote_id = __kmpc_impl_ffs(Liveness & lanemask_gt);
|
||||
size = __kmpc_impl_popc(Liveness);
|
||||
Liveness = __ACTIVEMASK();
|
||||
remote_id = __ffs(Liveness & lanemask_gt);
|
||||
size = __popc(Liveness);
|
||||
logical_lane_id /= 2;
|
||||
shflFct(reduce_data, /*LaneId =*/logical_lane_id,
|
||||
/*Offset=*/remote_id - 1 - physical_lane_id, /*AlgoVersion=*/2);
|
||||
@@ -81,8 +83,8 @@ int32_t __kmpc_nvptx_simd_reduce_nowait(int32_t global_tid, int32_t num_vars,
|
||||
size_t reduce_size, void *reduce_data,
|
||||
kmp_ShuffleReductFctPtr shflFct,
|
||||
kmp_InterWarpCopyFctPtr cpyFct) {
|
||||
__kmpc_impl_lanemask_t Liveness = __kmpc_impl_activemask();
|
||||
if (Liveness == __kmpc_impl_all_lanes) {
|
||||
uint32_t Liveness = __ACTIVEMASK();
|
||||
if (Liveness == 0xffffffff) {
|
||||
gpu_regular_warp_reduce(reduce_data, shflFct);
|
||||
return GetThreadIdInBlock() % WARPSIZE ==
|
||||
0; // Result on lane 0 of the simd warp.
|
||||
@@ -142,12 +144,12 @@ static int32_t nvptx_parallel_reduce_nowait(
|
||||
}
|
||||
return BlockThreadId == 0;
|
||||
#else
|
||||
__kmpc_impl_lanemask_t Liveness = __kmpc_impl_activemask();
|
||||
if (Liveness == __kmpc_impl_all_lanes) // Full warp
|
||||
uint32_t Liveness = __ACTIVEMASK();
|
||||
if (Liveness == 0xffffffff) // Full warp
|
||||
gpu_regular_warp_reduce(reduce_data, shflFct);
|
||||
else if (!(Liveness & (Liveness + 1))) // Partial warp but contiguous lanes
|
||||
gpu_irregular_warp_reduce(reduce_data, shflFct,
|
||||
/*LaneCount=*/__kmpc_impl_popc(Liveness),
|
||||
/*LaneCount=*/__popc(Liveness),
|
||||
/*LaneId=*/GetThreadIdInBlock() % WARPSIZE);
|
||||
else if (!isRuntimeUninitialized) // Dispersed lanes. Only threads in L2
|
||||
// parallel region may enter here; return
|
||||
@@ -317,12 +319,12 @@ static int32_t nvptx_teams_reduce_nowait(int32_t global_tid, int32_t num_vars,
|
||||
ldFct(reduce_data, scratchpad, i, NumTeams, /*Load and reduce*/ 1);
|
||||
|
||||
// Reduce across warps to the warp master.
|
||||
__kmpc_impl_lanemask_t Liveness = __kmpc_impl_activemask();
|
||||
if (Liveness == __kmpc_impl_all_lanes) // Full warp
|
||||
uint32_t Liveness = __ACTIVEMASK();
|
||||
if (Liveness == 0xffffffff) // Full warp
|
||||
gpu_regular_warp_reduce(reduce_data, shflFct);
|
||||
else // Partial warp but contiguous lanes
|
||||
gpu_irregular_warp_reduce(reduce_data, shflFct,
|
||||
/*LaneCount=*/__kmpc_impl_popc(Liveness),
|
||||
/*LaneCount=*/__popc(Liveness),
|
||||
/*LaneId=*/ThreadId % WARPSIZE);
|
||||
|
||||
// When we have more than [warpsize] number of threads
|
||||
|
||||
@@ -10,7 +10,6 @@
|
||||
//
|
||||
//===----------------------------------------------------------------------===//
|
||||
|
||||
#include "target_impl.h"
|
||||
////////////////////////////////////////////////////////////////////////////////
|
||||
// Execution Parameters
|
||||
////////////////////////////////////////////////////////////////////////////////
|
||||
@@ -66,8 +65,8 @@ INLINE int GetNumberOfProcsInDevice(bool isSPMDExecutionMode);
|
||||
INLINE int IsTeamMaster(int ompThreadId);
|
||||
|
||||
// Parallel level
|
||||
INLINE void IncParallelLevel(bool ActiveParallel, __kmpc_impl_lanemask_t Mask);
|
||||
INLINE void DecParallelLevel(bool ActiveParallel, __kmpc_impl_lanemask_t Mask);
|
||||
INLINE void IncParallelLevel(bool ActiveParallel);
|
||||
INLINE void DecParallelLevel(bool ActiveParallel);
|
||||
|
||||
////////////////////////////////////////////////////////////////////////////////
|
||||
// Memory
|
||||
|
||||
@@ -14,8 +14,6 @@
|
||||
// Execution Parameters
|
||||
////////////////////////////////////////////////////////////////////////////////
|
||||
|
||||
#include "target_impl.h"
|
||||
|
||||
INLINE void setExecutionParameters(ExecutionMode EMode, RuntimeMode RMode) {
|
||||
execution_param = EMode;
|
||||
execution_param |= RMode;
|
||||
@@ -203,28 +201,26 @@ INLINE int IsTeamMaster(int ompThreadId) { return (ompThreadId == 0); }
|
||||
////////////////////////////////////////////////////////////////////////////////
|
||||
// Parallel level
|
||||
|
||||
INLINE void IncParallelLevel(bool ActiveParallel, __kmpc_impl_lanemask_t Mask) {
|
||||
__kmpc_impl_syncwarp(Mask);
|
||||
__kmpc_impl_lanemask_t LaneMaskLt = __kmpc_impl_lanemask_lt();
|
||||
unsigned Rank = __kmpc_impl_popc(Mask & LaneMaskLt);
|
||||
if (Rank == 0) {
|
||||
INLINE void IncParallelLevel(bool ActiveParallel) {
|
||||
unsigned tnum = __ACTIVEMASK();
|
||||
int leader = __ffs(tnum) - 1;
|
||||
__SHFL_SYNC(tnum, leader, leader);
|
||||
if (GetLaneId() == leader) {
|
||||
parallelLevel[GetWarpId()] +=
|
||||
(1 + (ActiveParallel ? OMP_ACTIVE_PARALLEL_LEVEL : 0));
|
||||
__threadfence();
|
||||
}
|
||||
__kmpc_impl_syncwarp(Mask);
|
||||
__SHFL_SYNC(tnum, leader, leader);
|
||||
}
|
||||
|
||||
INLINE void DecParallelLevel(bool ActiveParallel, __kmpc_impl_lanemask_t Mask) {
|
||||
__kmpc_impl_syncwarp(Mask);
|
||||
__kmpc_impl_lanemask_t LaneMaskLt = __kmpc_impl_lanemask_lt();
|
||||
unsigned Rank = __kmpc_impl_popc(Mask & LaneMaskLt);
|
||||
if (Rank == 0) {
|
||||
INLINE void DecParallelLevel(bool ActiveParallel) {
|
||||
unsigned tnum = __ACTIVEMASK();
|
||||
int leader = __ffs(tnum) - 1;
|
||||
__SHFL_SYNC(tnum, leader, leader);
|
||||
if (GetLaneId() == leader) {
|
||||
parallelLevel[GetWarpId()] -=
|
||||
(1 + (ActiveParallel ? OMP_ACTIVE_PARALLEL_LEVEL : 0));
|
||||
__threadfence();
|
||||
}
|
||||
__kmpc_impl_syncwarp(Mask);
|
||||
__SHFL_SYNC(tnum, leader, leader);
|
||||
}
|
||||
|
||||
////////////////////////////////////////////////////////////////////////////////
|
||||
|
||||
@@ -11,7 +11,6 @@
|
||||
//===----------------------------------------------------------------------===//
|
||||
|
||||
#include "omptarget-nvptx.h"
|
||||
#include "target_impl.h"
|
||||
|
||||
////////////////////////////////////////////////////////////////////////////////
|
||||
// KMP Ordered calls
|
||||
@@ -63,9 +62,6 @@ EXTERN void __kmpc_barrier(kmp_Ident *loc_ref, int32_t tid) {
|
||||
// Barrier #1 is for synchronization among active threads.
|
||||
named_sync(L1_BARRIER, threads);
|
||||
}
|
||||
} else {
|
||||
// Still need to flush the memory per the standard.
|
||||
__kmpc_flush(loc_ref);
|
||||
} // numberOfActiveOMPThreads > 1
|
||||
PRINT0(LD_SYNC, "completed kmpc_barrier\n");
|
||||
}
|
||||
@@ -75,7 +71,8 @@ EXTERN void __kmpc_barrier(kmp_Ident *loc_ref, int32_t tid) {
|
||||
// parallel region and that all worker threads participate.
|
||||
EXTERN void __kmpc_barrier_simple_spmd(kmp_Ident *loc_ref, int32_t tid) {
|
||||
PRINT0(LD_SYNC, "call kmpc_barrier_simple_spmd\n");
|
||||
__kmpc_impl_syncthreads();
|
||||
// FIXME: use __syncthreads instead when the function copy is fixed in LLVM.
|
||||
__SYNCTHREADS();
|
||||
PRINT0(LD_SYNC, "completed kmpc_barrier_simple_spmd\n");
|
||||
}
|
||||
|
||||
@@ -140,16 +137,7 @@ EXTERN void __kmpc_flush(kmp_Ident *loc) {
|
||||
// Vote
|
||||
////////////////////////////////////////////////////////////////////////////////
|
||||
|
||||
EXTERN __kmpc_impl_lanemask_t __kmpc_warp_active_thread_mask() {
|
||||
EXTERN int32_t __kmpc_warp_active_thread_mask() {
|
||||
PRINT0(LD_IO, "call __kmpc_warp_active_thread_mask\n");
|
||||
return __kmpc_impl_activemask();
|
||||
}
|
||||
|
||||
////////////////////////////////////////////////////////////////////////////////
|
||||
// Syncwarp
|
||||
////////////////////////////////////////////////////////////////////////////////
|
||||
|
||||
EXTERN void __kmpc_syncwarp(__kmpc_impl_lanemask_t Mask) {
|
||||
PRINT0(LD_IO, "call __kmpc_syncwarp\n");
|
||||
__kmpc_impl_syncwarp(Mask);
|
||||
return __ACTIVEMASK();
|
||||
}
|
||||
|
||||
@@ -1,100 +0,0 @@
|
||||
//===------------ target_impl.h - NVPTX OpenMP GPU options ------- CUDA -*-===//
|
||||
//
|
||||
// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
|
||||
// See https://llvm.org/LICENSE.txt for license information.
|
||||
// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
|
||||
//
|
||||
//===----------------------------------------------------------------------===//
|
||||
//
|
||||
// Definitions of target specific functions
|
||||
//
|
||||
//===----------------------------------------------------------------------===//
|
||||
#ifndef _TARGET_IMPL_H_
|
||||
#define _TARGET_IMPL_H_
|
||||
|
||||
#include <stdint.h>
|
||||
|
||||
#include "option.h"
|
||||
|
||||
INLINE void __kmpc_impl_unpack(uint64_t val, uint32_t &lo, uint32_t &hi) {
|
||||
asm volatile("mov.b64 {%0,%1}, %2;" : "=r"(lo), "=r"(hi) : "l"(val));
|
||||
}
|
||||
|
||||
INLINE uint64_t __kmpc_impl_pack(uint32_t lo, uint32_t hi) {
|
||||
uint64_t val;
|
||||
asm volatile("mov.b64 %0, {%1,%2};" : "=l"(val) : "r"(lo), "r"(hi));
|
||||
return val;
|
||||
}
|
||||
|
||||
static const __kmpc_impl_lanemask_t __kmpc_impl_all_lanes =
|
||||
UINT32_C(0xffffffff);
|
||||
|
||||
INLINE __kmpc_impl_lanemask_t __kmpc_impl_lanemask_lt() {
|
||||
__kmpc_impl_lanemask_t res;
|
||||
asm("mov.u32 %0, %%lanemask_lt;" : "=r"(res));
|
||||
return res;
|
||||
}
|
||||
|
||||
INLINE __kmpc_impl_lanemask_t __kmpc_impl_lanemask_gt() {
|
||||
__kmpc_impl_lanemask_t res;
|
||||
asm("mov.u32 %0, %%lanemask_gt;" : "=r"(res));
|
||||
return res;
|
||||
}
|
||||
|
||||
INLINE uint32_t __kmpc_impl_ffs(uint32_t x) { return __ffs(x); }
|
||||
|
||||
INLINE uint32_t __kmpc_impl_popc(uint32_t x) { return __popc(x); }
|
||||
|
||||
#ifndef CUDA_VERSION
|
||||
#error CUDA_VERSION macro is undefined, something wrong with cuda.
|
||||
#endif
|
||||
|
||||
// In Cuda 9.0, __ballot(1) from Cuda 8.0 is replaced with __activemask().
|
||||
|
||||
INLINE __kmpc_impl_lanemask_t __kmpc_impl_activemask() {
|
||||
#if CUDA_VERSION >= 9000
|
||||
return __activemask();
|
||||
#else
|
||||
return __ballot(1);
|
||||
#endif
|
||||
}
|
||||
|
||||
// In Cuda 9.0, the *_sync() version takes an extra argument 'mask'.
|
||||
|
||||
INLINE int32_t __kmpc_impl_shfl_sync(__kmpc_impl_lanemask_t Mask, int32_t Var,
|
||||
int32_t SrcLane) {
|
||||
#if CUDA_VERSION >= 9000
|
||||
return __shfl_sync(Mask, Var, SrcLane);
|
||||
#else
|
||||
return __shfl(Var, SrcLane);
|
||||
#endif // CUDA_VERSION
|
||||
}
|
||||
|
||||
INLINE int32_t __kmpc_impl_shfl_down_sync(__kmpc_impl_lanemask_t Mask,
|
||||
int32_t Var, uint32_t Delta,
|
||||
int32_t Width) {
|
||||
#if CUDA_VERSION >= 9000
|
||||
return __shfl_down_sync(Mask, Var, Delta, Width);
|
||||
#else
|
||||
return __shfl_down(Var, Delta, Width);
|
||||
#endif // CUDA_VERSION
|
||||
}
|
||||
|
||||
INLINE void __kmpc_impl_syncthreads() {
|
||||
// Use original __syncthreads if compiled by nvcc or clang >= 9.0.
|
||||
#if !defined(__clang__) || __clang_major__ >= 9
|
||||
__syncthreads();
|
||||
#else
|
||||
asm volatile("bar.sync %0;" : : "r"(0) : "memory");
|
||||
#endif // __clang__
|
||||
}
|
||||
|
||||
INLINE void __kmpc_impl_syncwarp(__kmpc_impl_lanemask_t Mask) {
|
||||
#if CUDA_VERSION >= 9000
|
||||
__syncwarp(Mask);
|
||||
#else
|
||||
// In Cuda < 9.0 no need to sync threads in warps.
|
||||
#endif // CUDA_VERSION
|
||||
}
|
||||
|
||||
#endif
|
||||
@@ -1,37 +0,0 @@
|
||||
// RUN: %compile-run-and-check
|
||||
|
||||
#include <omp.h>
|
||||
#include <stdio.h>
|
||||
|
||||
int main(int argc, char *argv[]) {
|
||||
int data, out, flag = 0;
|
||||
#pragma omp target teams num_teams(2) map(tofrom \
|
||||
: out) map(to \
|
||||
: data, flag) \
|
||||
thread_limit(1)
|
||||
#pragma omp parallel num_threads(1)
|
||||
{
|
||||
if (omp_get_team_num() == 0) {
|
||||
/* Write to the data buffer that will be read by thread in team 1 */
|
||||
data = 42;
|
||||
/* Flush data to thread in team 1 */
|
||||
#pragma omp barrier
|
||||
/* Set flag to release thread in team 1 */
|
||||
#pragma omp atomic write
|
||||
flag = 1;
|
||||
} else if (omp_get_team_num() == 1) {
|
||||
/* Loop until we see the update to the flag */
|
||||
int val;
|
||||
do {
|
||||
#pragma omp atomic read
|
||||
val = flag;
|
||||
} while (val < 1);
|
||||
out = data;
|
||||
#pragma omp barrier
|
||||
}
|
||||
}
|
||||
// CHECK: out=42.
|
||||
/* Value of out will be 42 */
|
||||
printf("out=%d.\n", out);
|
||||
return !(out == 42);
|
||||
}
|
||||
@@ -135,17 +135,5 @@ int main(int argc, char *argv[]) {
|
||||
}
|
||||
}
|
||||
|
||||
// Check for paraller level in non-SPMD kernels.
|
||||
level = 0;
|
||||
#pragma omp target teams distribute num_teams(1) thread_limit(32) reduction(+:level)
|
||||
for (int i=0; i<5032; i+=32) {
|
||||
int ub = (i+32 > 5032) ? 5032 : i+32;
|
||||
#pragma omp parallel for schedule(dynamic)
|
||||
for (int j=i ; j < ub; j++) ;
|
||||
level += omp_get_level();
|
||||
}
|
||||
// CHECK: Integral level = 0.
|
||||
printf("Integral level = %d.\n", level);
|
||||
|
||||
return 0;
|
||||
}
|
||||
|
||||
@@ -47,8 +47,6 @@ enum tgt_map_type {
|
||||
OMP_TGT_MAPTYPE_LITERAL = 0x100,
|
||||
// mapping is implicit
|
||||
OMP_TGT_MAPTYPE_IMPLICIT = 0x200,
|
||||
// copy data to device
|
||||
OMP_TGT_MAPTYPE_CLOSE = 0x400,
|
||||
// member of struct, member given by [16 MSBs] - 1
|
||||
OMP_TGT_MAPTYPE_MEMBER_OF = 0xffff000000000000
|
||||
};
|
||||
|
||||
@@ -28,6 +28,10 @@ libomptarget_say("Building CUDA offloading plugin.")
|
||||
# Define the suffix for the runtime messaging dumps.
|
||||
add_definitions(-DTARGET_NAME=CUDA)
|
||||
|
||||
if(LIBOMPTARGET_CMAKE_BUILD_TYPE MATCHES debug)
|
||||
add_definitions(-DCUDA_ERROR_REPORT)
|
||||
endif()
|
||||
|
||||
include_directories(${LIBOMPTARGET_DEP_CUDA_INCLUDE_DIRS})
|
||||
include_directories(${LIBOMPTARGET_DEP_LIBELF_INCLUDE_DIRS})
|
||||
|
||||
|
||||
@@ -34,23 +34,25 @@ static int DebugLevel = 0;
|
||||
DEBUGP("Target " GETNAME(TARGET_NAME) " RTL", __VA_ARGS__); \
|
||||
} \
|
||||
} while (false)
|
||||
|
||||
// Utility for retrieving and printing CUDA error string.
|
||||
#define CUDA_ERR_STRING(err) \
|
||||
do { \
|
||||
if (DebugLevel > 0) { \
|
||||
const char *errStr; \
|
||||
cuGetErrorString(err, &errStr); \
|
||||
DEBUGP("Target " GETNAME(TARGET_NAME) " RTL", "CUDA error is: %s\n", errStr); \
|
||||
} \
|
||||
} while (false)
|
||||
#else // OMPTARGET_DEBUG
|
||||
#define DP(...) {}
|
||||
#define CUDA_ERR_STRING(err) {}
|
||||
#endif // OMPTARGET_DEBUG
|
||||
|
||||
#include "../../common/elf_common.c"
|
||||
|
||||
// Utility for retrieving and printing CUDA error string.
|
||||
#ifdef CUDA_ERROR_REPORT
|
||||
#define CUDA_ERR_STRING(err) \
|
||||
do { \
|
||||
const char *errStr; \
|
||||
cuGetErrorString(err, &errStr); \
|
||||
DP("CUDA error is: %s\n", errStr); \
|
||||
} while (0)
|
||||
#else
|
||||
#define CUDA_ERR_STRING(err) \
|
||||
{}
|
||||
#endif
|
||||
|
||||
/// Keep entries table per device.
|
||||
struct FuncOrGblEntryTy {
|
||||
__tgt_target_table Table;
|
||||
|
||||
@@ -50,7 +50,7 @@ static int DebugLevel = 0;
|
||||
#include "../../common/elf_common.c"
|
||||
|
||||
#define NUMBER_OF_DEVICES 4
|
||||
#define OFFLOADSECTIONNAME "omp_offloading_entries"
|
||||
#define OFFLOADSECTIONNAME ".omp_offloading.entries"
|
||||
|
||||
/// Array of Dynamic libraries loaded for this target.
|
||||
struct DynLibTy {
|
||||
|
||||
@@ -113,15 +113,7 @@ EXTERN int omp_target_is_present(void *ptr, int device_num) {
|
||||
|
||||
DeviceTy& Device = Devices[device_num];
|
||||
bool IsLast; // not used
|
||||
bool IsHostPtr;
|
||||
void *TgtPtr = Device.getTgtPtrBegin(ptr, 0, IsLast, false, IsHostPtr);
|
||||
int rc = (TgtPtr != NULL);
|
||||
// Under unified memory the host pointer can be returned by the
|
||||
// getTgtPtrBegin() function which means that there is no device
|
||||
// corresponding point for ptr. This function should return false
|
||||
// in that situation.
|
||||
if (RTLs.RequiresFlags & OMP_REQ_UNIFIED_SHARED_MEMORY)
|
||||
rc = !IsHostPtr;
|
||||
int rc = (Device.getTgtPtrBegin(ptr, 0, IsLast, false) != NULL);
|
||||
DP("Call to omp_target_is_present returns %d\n", rc);
|
||||
return rc;
|
||||
}
|
||||
|
||||
+15
-44
@@ -157,17 +157,12 @@ LookupResult DeviceTy::lookupMapping(void *HstPtrBegin, int64_t Size) {
|
||||
// If NULL is returned, then either data allocation failed or the user tried
|
||||
// to do an illegal mapping.
|
||||
void *DeviceTy::getOrAllocTgtPtr(void *HstPtrBegin, void *HstPtrBase,
|
||||
int64_t Size, bool &IsNew, bool &IsHostPtr, bool IsImplicit,
|
||||
bool UpdateRefCount, bool HasCloseModifier) {
|
||||
int64_t Size, bool &IsNew, bool IsImplicit, bool UpdateRefCount) {
|
||||
void *rc = NULL;
|
||||
IsHostPtr = false;
|
||||
DataMapMtx.lock();
|
||||
LookupResult lr = lookupMapping(HstPtrBegin, Size);
|
||||
|
||||
// Check if the pointer is contained.
|
||||
// If a variable is mapped to the device manually by the user - which would
|
||||
// lead to the IsContained flag to be true - then we must ensure that the
|
||||
// device address is returned even under unified memory conditions.
|
||||
if (lr.Flags.IsContained ||
|
||||
((lr.Flags.ExtendsBefore || lr.Flags.ExtendsAfter) && IsImplicit)) {
|
||||
auto &HT = *lr.Entry;
|
||||
@@ -188,28 +183,15 @@ void *DeviceTy::getOrAllocTgtPtr(void *HstPtrBegin, void *HstPtrBase,
|
||||
// Explicit extension of mapped data - not allowed.
|
||||
DP("Explicit extension of mapping is not allowed.\n");
|
||||
} else if (Size) {
|
||||
// If unified shared memory is active, implicitly mapped variables that are not
|
||||
// privatized use host address. Any explicitly mapped variables also use
|
||||
// host address where correctness is not impeded. In all other cases
|
||||
// maps are respected.
|
||||
// In addition to the mapping rules above, the close map
|
||||
// modifier forces the mapping of the variable to the device.
|
||||
if (RTLs.RequiresFlags & OMP_REQ_UNIFIED_SHARED_MEMORY && !HasCloseModifier) {
|
||||
DP("Return HstPtrBegin " DPxMOD " Size=%ld RefCount=%s\n",
|
||||
DPxPTR((uintptr_t)HstPtrBegin), Size, (UpdateRefCount ? " updated" : ""));
|
||||
IsHostPtr = true;
|
||||
rc = HstPtrBegin;
|
||||
} else {
|
||||
// If it is not contained and Size > 0 we should create a new entry for it.
|
||||
IsNew = true;
|
||||
uintptr_t tp = (uintptr_t)RTL->data_alloc(RTLDeviceID, Size, HstPtrBegin);
|
||||
DP("Creating new map entry: HstBase=" DPxMOD ", HstBegin=" DPxMOD ", "
|
||||
"HstEnd=" DPxMOD ", TgtBegin=" DPxMOD "\n", DPxPTR(HstPtrBase),
|
||||
DPxPTR(HstPtrBegin), DPxPTR((uintptr_t)HstPtrBegin + Size), DPxPTR(tp));
|
||||
HostDataToTargetMap.push_front(HostDataToTargetTy((uintptr_t)HstPtrBase,
|
||||
(uintptr_t)HstPtrBegin, (uintptr_t)HstPtrBegin + Size, tp));
|
||||
rc = (void *)tp;
|
||||
}
|
||||
// If it is not contained and Size > 0 we should create a new entry for it.
|
||||
IsNew = true;
|
||||
uintptr_t tp = (uintptr_t)RTL->data_alloc(RTLDeviceID, Size, HstPtrBegin);
|
||||
DP("Creating new map entry: HstBase=" DPxMOD ", HstBegin=" DPxMOD ", "
|
||||
"HstEnd=" DPxMOD ", TgtBegin=" DPxMOD "\n", DPxPTR(HstPtrBase),
|
||||
DPxPTR(HstPtrBegin), DPxPTR((uintptr_t)HstPtrBegin + Size), DPxPTR(tp));
|
||||
HostDataToTargetMap.push_front(HostDataToTargetTy((uintptr_t)HstPtrBase,
|
||||
(uintptr_t)HstPtrBegin, (uintptr_t)HstPtrBegin + Size, tp));
|
||||
rc = (void *)tp;
|
||||
}
|
||||
|
||||
DataMapMtx.unlock();
|
||||
@@ -220,10 +202,8 @@ void *DeviceTy::getOrAllocTgtPtr(void *HstPtrBegin, void *HstPtrBase,
|
||||
// Return the target pointer begin (where the data will be moved).
|
||||
// Decrement the reference counter if called from target_data_end.
|
||||
void *DeviceTy::getTgtPtrBegin(void *HstPtrBegin, int64_t Size, bool &IsLast,
|
||||
bool UpdateRefCount, bool &IsHostPtr) {
|
||||
bool UpdateRefCount) {
|
||||
void *rc = NULL;
|
||||
IsHostPtr = false;
|
||||
IsLast = false;
|
||||
DataMapMtx.lock();
|
||||
LookupResult lr = lookupMapping(HstPtrBegin, Size);
|
||||
|
||||
@@ -241,14 +221,8 @@ void *DeviceTy::getTgtPtrBegin(void *HstPtrBegin, int64_t Size, bool &IsLast,
|
||||
(CONSIDERED_INF(HT.RefCount)) ? "INF" :
|
||||
std::to_string(HT.RefCount).c_str());
|
||||
rc = (void *)tp;
|
||||
} else if (RTLs.RequiresFlags & OMP_REQ_UNIFIED_SHARED_MEMORY) {
|
||||
// If the value isn't found in the mapping and unified shared memory
|
||||
// is on then it means we have stumbled upon a value which we need to
|
||||
// use directly from the host.
|
||||
DP("Get HstPtrBegin " DPxMOD " Size=%ld RefCount=%s\n",
|
||||
DPxPTR((uintptr_t)HstPtrBegin), Size, (UpdateRefCount ? " updated" : ""));
|
||||
IsHostPtr = true;
|
||||
rc = HstPtrBegin;
|
||||
} else {
|
||||
IsLast = false;
|
||||
}
|
||||
|
||||
DataMapMtx.unlock();
|
||||
@@ -269,10 +243,7 @@ void *DeviceTy::getTgtPtrBegin(void *HstPtrBegin, int64_t Size) {
|
||||
return NULL;
|
||||
}
|
||||
|
||||
int DeviceTy::deallocTgtPtr(void *HstPtrBegin, int64_t Size, bool ForceDelete,
|
||||
bool HasCloseModifier) {
|
||||
if (RTLs.RequiresFlags & OMP_REQ_UNIFIED_SHARED_MEMORY && !HasCloseModifier)
|
||||
return OFFLOAD_SUCCESS;
|
||||
int DeviceTy::deallocTgtPtr(void *HstPtrBegin, int64_t Size, bool ForceDelete) {
|
||||
// Check if the pointer is contained in any sub-nodes.
|
||||
int rc;
|
||||
DataMapMtx.lock();
|
||||
@@ -306,7 +277,7 @@ int DeviceTy::deallocTgtPtr(void *HstPtrBegin, int64_t Size, bool ForceDelete,
|
||||
void DeviceTy::init() {
|
||||
// Make call to init_requires if it exists for this plugin.
|
||||
if (RTL->init_requires)
|
||||
RTL->init_requires(RTLs.RequiresFlags);
|
||||
RTL->init_requires(RTLRequiresFlags);
|
||||
int32_t rc = RTL->init_device(RTLDeviceID);
|
||||
if (rc == OFFLOAD_SUCCESS) {
|
||||
IsInit = true;
|
||||
|
||||
@@ -100,10 +100,13 @@ struct DeviceTy {
|
||||
// moved into the target task in libomp.
|
||||
std::map<int32_t, uint64_t> LoopTripCnt;
|
||||
|
||||
int64_t RTLRequiresFlags;
|
||||
|
||||
DeviceTy(RTLInfoTy *RTL)
|
||||
: DeviceID(-1), RTL(RTL), RTLDeviceID(-1), IsInit(false), InitFlag(),
|
||||
HasPendingGlobals(false), HostDataToTargetMap(), PendingCtorsDtors(),
|
||||
ShadowPtrMap(), DataMapMtx(), PendingGlobalsMtx(), ShadowMtx() {}
|
||||
HasPendingGlobals(false), HostDataToTargetMap(),
|
||||
PendingCtorsDtors(), ShadowPtrMap(), DataMapMtx(), PendingGlobalsMtx(),
|
||||
ShadowMtx(), RTLRequiresFlags(0) {}
|
||||
|
||||
// The existence of mutexes makes DeviceTy non-copyable. We need to
|
||||
// provide a copy constructor and an assignment operator explicitly.
|
||||
@@ -112,8 +115,9 @@ struct DeviceTy {
|
||||
IsInit(d.IsInit), InitFlag(), HasPendingGlobals(d.HasPendingGlobals),
|
||||
HostDataToTargetMap(d.HostDataToTargetMap),
|
||||
PendingCtorsDtors(d.PendingCtorsDtors), ShadowPtrMap(d.ShadowPtrMap),
|
||||
DataMapMtx(), PendingGlobalsMtx(), ShadowMtx(),
|
||||
LoopTripCnt(d.LoopTripCnt) {}
|
||||
DataMapMtx(), PendingGlobalsMtx(),
|
||||
ShadowMtx(), LoopTripCnt(d.LoopTripCnt),
|
||||
RTLRequiresFlags(d.RTLRequiresFlags) {}
|
||||
|
||||
DeviceTy& operator=(const DeviceTy &d) {
|
||||
DeviceID = d.DeviceID;
|
||||
@@ -125,6 +129,7 @@ struct DeviceTy {
|
||||
PendingCtorsDtors = d.PendingCtorsDtors;
|
||||
ShadowPtrMap = d.ShadowPtrMap;
|
||||
LoopTripCnt = d.LoopTripCnt;
|
||||
RTLRequiresFlags = d.RTLRequiresFlags;
|
||||
|
||||
return *this;
|
||||
}
|
||||
@@ -132,13 +137,11 @@ struct DeviceTy {
|
||||
long getMapEntryRefCnt(void *HstPtrBegin);
|
||||
LookupResult lookupMapping(void *HstPtrBegin, int64_t Size);
|
||||
void *getOrAllocTgtPtr(void *HstPtrBegin, void *HstPtrBase, int64_t Size,
|
||||
bool &IsNew, bool &IsHostPtr, bool IsImplicit, bool UpdateRefCount = true,
|
||||
bool HasCloseModifier = false);
|
||||
bool &IsNew, bool IsImplicit, bool UpdateRefCount = true);
|
||||
void *getTgtPtrBegin(void *HstPtrBegin, int64_t Size);
|
||||
void *getTgtPtrBegin(void *HstPtrBegin, int64_t Size, bool &IsLast,
|
||||
bool UpdateRefCount, bool &IsHostPtr);
|
||||
int deallocTgtPtr(void *TgtPtrBegin, int64_t Size, bool ForceDelete,
|
||||
bool HasCloseModifier = false);
|
||||
bool UpdateRefCount);
|
||||
int deallocTgtPtr(void *TgtPtrBegin, int64_t Size, bool ForceDelete);
|
||||
int associatePtr(void *HstPtrBegin, void *TgtPtrBegin, int64_t Size);
|
||||
int disassociatePtr(void *HstPtrBegin);
|
||||
|
||||
|
||||
@@ -13,8 +13,6 @@ VERS1.0 {
|
||||
__tgt_target_data_update_nowait;
|
||||
__tgt_target_nowait;
|
||||
__tgt_target_teams_nowait;
|
||||
__tgt_mapper_num_components;
|
||||
__tgt_push_mapper_component;
|
||||
omp_get_num_devices;
|
||||
omp_get_initial_device;
|
||||
omp_target_alloc;
|
||||
|
||||
@@ -304,33 +304,8 @@ EXTERN int __tgt_target_teams_nowait(int64_t device_id, void *host_ptr,
|
||||
arg_sizes, arg_types, team_num, thread_limit);
|
||||
}
|
||||
|
||||
// Get the current number of components for a user-defined mapper.
|
||||
EXTERN int64_t __tgt_mapper_num_components(void *rt_mapper_handle) {
|
||||
auto *MapperComponentsPtr = (struct MapperComponentsTy *)rt_mapper_handle;
|
||||
int64_t size = MapperComponentsPtr->Components.size();
|
||||
DP("__tgt_mapper_num_components(Handle=" DPxMOD ") returns %" PRId64 "\n",
|
||||
DPxPTR(rt_mapper_handle), size);
|
||||
return size;
|
||||
}
|
||||
|
||||
// Push back one component for a user-defined mapper.
|
||||
EXTERN void __tgt_push_mapper_component(void *rt_mapper_handle, void *base,
|
||||
void *begin, int64_t size,
|
||||
int64_t type) {
|
||||
DP("__tgt_push_mapper_component(Handle=" DPxMOD
|
||||
") adds an entry (Base=" DPxMOD ", Begin=" DPxMOD ", Size=%" PRId64
|
||||
", Type=0x%" PRIx64 ").\n",
|
||||
DPxPTR(rt_mapper_handle), DPxPTR(base), DPxPTR(begin), size, type);
|
||||
auto *MapperComponentsPtr = (struct MapperComponentsTy *)rt_mapper_handle;
|
||||
MapperComponentsPtr->Components.push_back(
|
||||
MapComponentInfoTy(base, begin, size, type));
|
||||
}
|
||||
|
||||
EXTERN void __kmpc_push_target_tripcount(int64_t device_id,
|
||||
uint64_t loop_tripcount) {
|
||||
if (IsOffloadDisabled())
|
||||
return;
|
||||
|
||||
if (device_id == OFFLOAD_DEVICE_DEFAULT) {
|
||||
device_id = omp_get_default_device();
|
||||
}
|
||||
|
||||
@@ -242,11 +242,7 @@ int target_data_begin(DeviceTy &Device, int32_t arg_num,
|
||||
// Address of pointer on the host and device, respectively.
|
||||
void *Pointer_HstPtrBegin, *Pointer_TgtPtrBegin;
|
||||
bool IsNew, Pointer_IsNew;
|
||||
bool IsHostPtr = false;
|
||||
bool IsImplicit = arg_types[i] & OMP_TGT_MAPTYPE_IMPLICIT;
|
||||
// Force the creation of a device side copy of the data when:
|
||||
// a close map modifier was associated with a map that contained a to.
|
||||
bool HasCloseModifier = arg_types[i] & OMP_TGT_MAPTYPE_CLOSE;
|
||||
// UpdateRef is based on MEMBER_OF instead of TARGET_PARAM because if we
|
||||
// have reached this point via __tgt_target_data_begin and not __tgt_target
|
||||
// then no argument is marked as TARGET_PARAM ("omp target data map" is not
|
||||
@@ -257,8 +253,7 @@ int target_data_begin(DeviceTy &Device, int32_t arg_num,
|
||||
DP("Has a pointer entry: \n");
|
||||
// base is address of pointer.
|
||||
Pointer_TgtPtrBegin = Device.getOrAllocTgtPtr(HstPtrBase, HstPtrBase,
|
||||
sizeof(void *), Pointer_IsNew, IsHostPtr, IsImplicit, UpdateRef,
|
||||
HasCloseModifier);
|
||||
sizeof(void *), Pointer_IsNew, IsImplicit, UpdateRef);
|
||||
if (!Pointer_TgtPtrBegin) {
|
||||
DP("Call to getOrAllocTgtPtr returned null pointer (device failure or "
|
||||
"illegal mapping).\n");
|
||||
@@ -274,7 +269,7 @@ int target_data_begin(DeviceTy &Device, int32_t arg_num,
|
||||
}
|
||||
|
||||
void *TgtPtrBegin = Device.getOrAllocTgtPtr(HstPtrBegin, HstPtrBase,
|
||||
data_size, IsNew, IsHostPtr, IsImplicit, UpdateRef, HasCloseModifier);
|
||||
data_size, IsNew, IsImplicit, UpdateRef);
|
||||
if (!TgtPtrBegin && data_size) {
|
||||
// If data_size==0, then the argument could be a zero-length pointer to
|
||||
// NULL, so getOrAlloc() returning NULL is not an error.
|
||||
@@ -294,22 +289,19 @@ int target_data_begin(DeviceTy &Device, int32_t arg_num,
|
||||
|
||||
if (arg_types[i] & OMP_TGT_MAPTYPE_TO) {
|
||||
bool copy = false;
|
||||
if (!(RTLs.RequiresFlags & OMP_REQ_UNIFIED_SHARED_MEMORY) ||
|
||||
HasCloseModifier) {
|
||||
if (IsNew || (arg_types[i] & OMP_TGT_MAPTYPE_ALWAYS)) {
|
||||
if (IsNew || (arg_types[i] & OMP_TGT_MAPTYPE_ALWAYS)) {
|
||||
copy = true;
|
||||
} else if (arg_types[i] & OMP_TGT_MAPTYPE_MEMBER_OF) {
|
||||
// Copy data only if the "parent" struct has RefCount==1.
|
||||
int32_t parent_idx = member_of(arg_types[i]);
|
||||
long parent_rc = Device.getMapEntryRefCnt(args[parent_idx]);
|
||||
assert(parent_rc > 0 && "parent struct not found");
|
||||
if (parent_rc == 1) {
|
||||
copy = true;
|
||||
} else if (arg_types[i] & OMP_TGT_MAPTYPE_MEMBER_OF) {
|
||||
// Copy data only if the "parent" struct has RefCount==1.
|
||||
int32_t parent_idx = member_of(arg_types[i]);
|
||||
long parent_rc = Device.getMapEntryRefCnt(args[parent_idx]);
|
||||
assert(parent_rc > 0 && "parent struct not found");
|
||||
if (parent_rc == 1) {
|
||||
copy = true;
|
||||
}
|
||||
}
|
||||
}
|
||||
|
||||
if (copy && !IsHostPtr) {
|
||||
if (copy) {
|
||||
DP("Moving %" PRId64 " bytes (hst:" DPxMOD ") -> (tgt:" DPxMOD ")\n",
|
||||
data_size, DPxPTR(HstPtrBegin), DPxPTR(TgtPtrBegin));
|
||||
int rt = Device.data_submit(TgtPtrBegin, HstPtrBegin, data_size);
|
||||
@@ -320,7 +312,7 @@ int target_data_begin(DeviceTy &Device, int32_t arg_num,
|
||||
}
|
||||
}
|
||||
|
||||
if (arg_types[i] & OMP_TGT_MAPTYPE_PTR_AND_OBJ && !IsHostPtr) {
|
||||
if (arg_types[i] & OMP_TGT_MAPTYPE_PTR_AND_OBJ) {
|
||||
DP("Update pointer (" DPxMOD ") -> [" DPxMOD "]\n",
|
||||
DPxPTR(Pointer_TgtPtrBegin), DPxPTR(TgtPtrBegin));
|
||||
uint64_t Delta = (uint64_t)HstPtrBegin - (uint64_t)HstPtrBase;
|
||||
@@ -371,15 +363,14 @@ int target_data_end(DeviceTy &Device, int32_t arg_num, void **args_base,
|
||||
}
|
||||
}
|
||||
|
||||
bool IsLast, IsHostPtr;
|
||||
bool IsLast;
|
||||
bool UpdateRef = !(arg_types[i] & OMP_TGT_MAPTYPE_MEMBER_OF) ||
|
||||
(arg_types[i] & OMP_TGT_MAPTYPE_PTR_AND_OBJ);
|
||||
bool ForceDelete = arg_types[i] & OMP_TGT_MAPTYPE_DELETE;
|
||||
bool HasCloseModifier = arg_types[i] & OMP_TGT_MAPTYPE_CLOSE;
|
||||
|
||||
// If PTR_AND_OBJ, HstPtrBegin is address of pointee
|
||||
void *TgtPtrBegin = Device.getTgtPtrBegin(HstPtrBegin, data_size, IsLast,
|
||||
UpdateRef, IsHostPtr);
|
||||
UpdateRef);
|
||||
DP("There are %" PRId64 " bytes allocated at target address " DPxMOD
|
||||
" - is%s last\n", data_size, DPxPTR(TgtPtrBegin),
|
||||
(IsLast ? "" : " not"));
|
||||
@@ -396,23 +387,18 @@ int target_data_end(DeviceTy &Device, int32_t arg_num, void **args_base,
|
||||
if (arg_types[i] & OMP_TGT_MAPTYPE_FROM) {
|
||||
bool Always = arg_types[i] & OMP_TGT_MAPTYPE_ALWAYS;
|
||||
bool CopyMember = false;
|
||||
if (!(RTLs.RequiresFlags & OMP_REQ_UNIFIED_SHARED_MEMORY) ||
|
||||
HasCloseModifier) {
|
||||
if ((arg_types[i] & OMP_TGT_MAPTYPE_MEMBER_OF) &&
|
||||
!(arg_types[i] & OMP_TGT_MAPTYPE_PTR_AND_OBJ)) {
|
||||
// Copy data only if the "parent" struct has RefCount==1.
|
||||
int32_t parent_idx = member_of(arg_types[i]);
|
||||
long parent_rc = Device.getMapEntryRefCnt(args[parent_idx]);
|
||||
assert(parent_rc > 0 && "parent struct not found");
|
||||
if (parent_rc == 1) {
|
||||
CopyMember = true;
|
||||
}
|
||||
if ((arg_types[i] & OMP_TGT_MAPTYPE_MEMBER_OF) &&
|
||||
!(arg_types[i] & OMP_TGT_MAPTYPE_PTR_AND_OBJ)) {
|
||||
// Copy data only if the "parent" struct has RefCount==1.
|
||||
int32_t parent_idx = member_of(arg_types[i]);
|
||||
long parent_rc = Device.getMapEntryRefCnt(args[parent_idx]);
|
||||
assert(parent_rc > 0 && "parent struct not found");
|
||||
if (parent_rc == 1) {
|
||||
CopyMember = true;
|
||||
}
|
||||
}
|
||||
|
||||
if ((DelEntry || Always || CopyMember) &&
|
||||
!(RTLs.RequiresFlags & OMP_REQ_UNIFIED_SHARED_MEMORY &&
|
||||
TgtPtrBegin == HstPtrBegin)) {
|
||||
if (DelEntry || Always || CopyMember) {
|
||||
DP("Moving %" PRId64 " bytes (tgt:" DPxMOD ") -> (hst:" DPxMOD ")\n",
|
||||
data_size, DPxPTR(TgtPtrBegin), DPxPTR(HstPtrBegin));
|
||||
int rt = Device.data_retrieve(HstPtrBegin, TgtPtrBegin, data_size);
|
||||
@@ -462,8 +448,7 @@ int target_data_end(DeviceTy &Device, int32_t arg_num, void **args_base,
|
||||
|
||||
// Deallocate map
|
||||
if (DelEntry) {
|
||||
int rt = Device.deallocTgtPtr(HstPtrBegin, data_size, ForceDelete,
|
||||
HasCloseModifier);
|
||||
int rt = Device.deallocTgtPtr(HstPtrBegin, data_size, ForceDelete);
|
||||
if (rt != OFFLOAD_SUCCESS) {
|
||||
DP("Deallocating data from device failed.\n");
|
||||
return OFFLOAD_FAIL;
|
||||
@@ -486,21 +471,14 @@ int target_data_update(DeviceTy &Device, int32_t arg_num,
|
||||
|
||||
void *HstPtrBegin = args[i];
|
||||
int64_t MapSize = arg_sizes[i];
|
||||
bool IsLast, IsHostPtr;
|
||||
bool IsLast;
|
||||
void *TgtPtrBegin = Device.getTgtPtrBegin(HstPtrBegin, MapSize, IsLast,
|
||||
false, IsHostPtr);
|
||||
false);
|
||||
if (!TgtPtrBegin) {
|
||||
DP("hst data:" DPxMOD " not found, becomes a noop\n", DPxPTR(HstPtrBegin));
|
||||
continue;
|
||||
}
|
||||
|
||||
if (RTLs.RequiresFlags & OMP_REQ_UNIFIED_SHARED_MEMORY &&
|
||||
TgtPtrBegin == HstPtrBegin) {
|
||||
DP("hst data:" DPxMOD " unified and shared, becomes a noop\n",
|
||||
DPxPTR(HstPtrBegin));
|
||||
continue;
|
||||
}
|
||||
|
||||
if (arg_types[i] & OMP_TGT_MAPTYPE_FROM) {
|
||||
DP("Moving %" PRId64 " bytes (tgt:" DPxMOD ") -> (hst:" DPxMOD ")\n",
|
||||
arg_sizes[i], DPxPTR(TgtPtrBegin), DPxPTR(HstPtrBegin));
|
||||
@@ -536,7 +514,6 @@ int target_data_update(DeviceTy &Device, int32_t arg_num,
|
||||
DP("Copying data to device failed.\n");
|
||||
return OFFLOAD_FAIL;
|
||||
}
|
||||
|
||||
uintptr_t lb = (uintptr_t) HstPtrBegin;
|
||||
uintptr_t ub = (uintptr_t) HstPtrBegin + MapSize;
|
||||
Device.ShadowMtx.lock();
|
||||
@@ -663,26 +640,19 @@ int target(int64_t device_id, void *host_ptr, int32_t arg_num,
|
||||
void *HstPtrVal = args[i];
|
||||
void *HstPtrBegin = args_base[i];
|
||||
void *HstPtrBase = args[idx];
|
||||
bool IsLast, IsHostPtr; // unused.
|
||||
bool IsLast; // unused.
|
||||
void *TgtPtrBase =
|
||||
(void *)((intptr_t)tgt_args[tgtIdx] + tgt_offsets[tgtIdx]);
|
||||
DP("Parent lambda base " DPxMOD "\n", DPxPTR(TgtPtrBase));
|
||||
uint64_t Delta = (uint64_t)HstPtrBegin - (uint64_t)HstPtrBase;
|
||||
void *TgtPtrBegin = (void *)((uintptr_t)TgtPtrBase + Delta);
|
||||
void *Pointer_TgtPtrBegin =
|
||||
Device.getTgtPtrBegin(HstPtrVal, arg_sizes[i], IsLast, false,
|
||||
IsHostPtr);
|
||||
Device.getTgtPtrBegin(HstPtrVal, arg_sizes[i], IsLast, false);
|
||||
if (!Pointer_TgtPtrBegin) {
|
||||
DP("No lambda captured variable mapped (" DPxMOD ") - ignored\n",
|
||||
DPxPTR(HstPtrVal));
|
||||
continue;
|
||||
}
|
||||
if (RTLs.RequiresFlags & OMP_REQ_UNIFIED_SHARED_MEMORY &&
|
||||
TgtPtrBegin == HstPtrBegin) {
|
||||
DP("Unified memory is active, no need to map lambda captured"
|
||||
"variable (" DPxMOD ")\n", DPxPTR(HstPtrVal));
|
||||
continue;
|
||||
}
|
||||
DP("Update lambda reference (" DPxMOD ") -> [" DPxMOD "]\n",
|
||||
DPxPTR(Pointer_TgtPtrBegin), DPxPTR(TgtPtrBegin));
|
||||
int rt = Device.data_submit(TgtPtrBegin, &Pointer_TgtPtrBegin,
|
||||
@@ -698,7 +668,7 @@ int target(int64_t device_id, void *host_ptr, int32_t arg_num,
|
||||
void *HstPtrBase = args_base[i];
|
||||
void *TgtPtrBegin;
|
||||
ptrdiff_t TgtBaseOffset;
|
||||
bool IsLast, IsHostPtr; // unused.
|
||||
bool IsLast; // unused.
|
||||
if (arg_types[i] & OMP_TGT_MAPTYPE_LITERAL) {
|
||||
DP("Forwarding first-private value " DPxMOD " to the target construct\n",
|
||||
DPxPTR(HstPtrBase));
|
||||
@@ -735,14 +705,14 @@ int target(int64_t device_id, void *host_ptr, int32_t arg_num,
|
||||
}
|
||||
} else if (arg_types[i] & OMP_TGT_MAPTYPE_PTR_AND_OBJ) {
|
||||
TgtPtrBegin = Device.getTgtPtrBegin(HstPtrBase, sizeof(void *), IsLast,
|
||||
false, IsHostPtr);
|
||||
false);
|
||||
TgtBaseOffset = 0; // no offset for ptrs.
|
||||
DP("Obtained target argument " DPxMOD " from host pointer " DPxMOD " to "
|
||||
"object " DPxMOD "\n", DPxPTR(TgtPtrBegin), DPxPTR(HstPtrBase),
|
||||
DPxPTR(HstPtrBase));
|
||||
} else {
|
||||
TgtPtrBegin = Device.getTgtPtrBegin(HstPtrBegin, arg_sizes[i], IsLast,
|
||||
false, IsHostPtr);
|
||||
false);
|
||||
TgtBaseOffset = (intptr_t)HstPtrBase - (intptr_t)HstPtrBegin;
|
||||
#ifdef OMPTARGET_DEBUG
|
||||
void *TgtPtrBase = (void *)((intptr_t)TgtPtrBegin + TgtBaseOffset);
|
||||
|
||||
@@ -41,24 +41,6 @@ enum kmp_target_offload_kind {
|
||||
typedef enum kmp_target_offload_kind kmp_target_offload_kind_t;
|
||||
extern kmp_target_offload_kind_t TargetOffloadPolicy;
|
||||
|
||||
// This structure stores information of a mapped memory region.
|
||||
struct MapComponentInfoTy {
|
||||
void *Base;
|
||||
void *Begin;
|
||||
int64_t Size;
|
||||
int64_t Type;
|
||||
MapComponentInfoTy() = default;
|
||||
MapComponentInfoTy(void *Base, void *Begin, int64_t Size, int64_t Type)
|
||||
: Base(Base), Begin(Begin), Size(Size), Type(Type) {}
|
||||
};
|
||||
|
||||
// This structure stores all components of a user-defined mapper. The number of
|
||||
// components are dynamically decided, so we utilize C++ STL vector
|
||||
// implementation here.
|
||||
struct MapperComponentsTy {
|
||||
std::vector<MapComponentInfoTy> Components;
|
||||
};
|
||||
|
||||
////////////////////////////////////////////////////////////////////////////////
|
||||
// implemtation for fatal messages
|
||||
////////////////////////////////////////////////////////////////////////////////
|
||||
|
||||
@@ -266,6 +266,8 @@ void RTLsTy::RegisterLib(__tgt_bin_desc *desc) {
|
||||
Devices[start + device_id].DeviceID = start + device_id;
|
||||
// RTL local device ID
|
||||
Devices[start + device_id].RTLDeviceID = device_id;
|
||||
// RTL requires flags
|
||||
Devices[start + device_id].RTLRequiresFlags = RequiresFlags;
|
||||
}
|
||||
|
||||
// Initialize the index of this RTL and save it in the used RTLs.
|
||||
|
||||
@@ -1,47 +0,0 @@
|
||||
// RUN: %libomptarget-compilexx-run-and-check-aarch64-unknown-linux-gnu
|
||||
// RUN: %libomptarget-compilexx-run-and-check-powerpc64-ibm-linux-gnu
|
||||
// RUN: %libomptarget-compilexx-run-and-check-powerpc64le-ibm-linux-gnu
|
||||
// RUN: %libomptarget-compilexx-run-and-check-x86_64-pc-linux-gnu
|
||||
|
||||
#include <cstdio>
|
||||
#include <cstdlib>
|
||||
#include <vector>
|
||||
|
||||
// Data structure definitions copied from OpenMP RTL.
|
||||
struct MapComponentInfoTy {
|
||||
void *Base;
|
||||
void *Begin;
|
||||
int64_t Size;
|
||||
int64_t Type;
|
||||
MapComponentInfoTy() = default;
|
||||
MapComponentInfoTy(void *Base, void *Begin, int64_t Size, int64_t Type)
|
||||
: Base(Base), Begin(Begin), Size(Size), Type(Type) {}
|
||||
};
|
||||
|
||||
struct MapperComponentsTy {
|
||||
std::vector<MapComponentInfoTy> Components;
|
||||
};
|
||||
|
||||
// OpenMP RTL interfaces
|
||||
#ifdef __cplusplus
|
||||
extern "C" {
|
||||
#endif
|
||||
int64_t __tgt_mapper_num_components(void *rt_mapper_handle);
|
||||
void __tgt_push_mapper_component(void *rt_mapper_handle, void *base,
|
||||
void *begin, int64_t size, int64_t type);
|
||||
#ifdef __cplusplus
|
||||
}
|
||||
#endif
|
||||
|
||||
int main(int argc, char *argv[]) {
|
||||
MapperComponentsTy MC;
|
||||
void *base, *begin;
|
||||
int64_t size, type;
|
||||
// Push 2 elements into MC.
|
||||
__tgt_push_mapper_component((void *)&MC, base, begin, size, type);
|
||||
__tgt_push_mapper_component((void *)&MC, base, begin, size, type);
|
||||
int64_t num = __tgt_mapper_num_components((void *)&MC);
|
||||
// CHECK: num=2
|
||||
printf("num=%lld\n", num);
|
||||
return 0;
|
||||
}
|
||||
@@ -43,4 +43,4 @@ int main() {
|
||||
{}
|
||||
|
||||
return 0;
|
||||
}
|
||||
}
|
||||
@@ -1,164 +0,0 @@
|
||||
// RUN: %libomptarget-compile-run-and-check-aarch64-unknown-linux-gnu
|
||||
// RUN: %libomptarget-compile-run-and-check-powerpc64-ibm-linux-gnu
|
||||
// RUN: %libomptarget-compile-run-and-check-powerpc64le-ibm-linux-gnu
|
||||
// RUN: %libomptarget-compile-run-and-check-x86_64-pc-linux-gnu
|
||||
|
||||
#include <stdio.h>
|
||||
#include <omp.h>
|
||||
|
||||
// ---------------------------------------------------------------------------
|
||||
// Various definitions copied from OpenMP RTL
|
||||
|
||||
extern void __tgt_register_requires(int64_t);
|
||||
|
||||
// End of definitions copied from OpenMP RTL.
|
||||
// ---------------------------------------------------------------------------
|
||||
|
||||
#pragma omp requires unified_shared_memory
|
||||
|
||||
#define N 1024
|
||||
|
||||
void init(int A[], int B[], int C[]) {
|
||||
for (int i = 0; i < N; ++i) {
|
||||
A[i] = 0;
|
||||
B[i] = 1;
|
||||
C[i] = i;
|
||||
}
|
||||
}
|
||||
|
||||
int main(int argc, char *argv[]) {
|
||||
const int device = omp_get_default_device();
|
||||
|
||||
// Manual registration of requires flags for Clang versions
|
||||
// that do not support requires.
|
||||
__tgt_register_requires(8);
|
||||
|
||||
// CHECK: Initial device: -10
|
||||
printf("Initial device: %d\n", omp_get_initial_device());
|
||||
|
||||
//
|
||||
// Target alloc & target memcpy
|
||||
//
|
||||
int A[N], B[N], C[N];
|
||||
|
||||
// Init
|
||||
init(A, B, C);
|
||||
|
||||
int *pA, *pB, *pC;
|
||||
|
||||
// map ptrs
|
||||
pA = &A[0];
|
||||
pB = &B[0];
|
||||
pC = &C[0];
|
||||
|
||||
int *d_A = (int *)omp_target_alloc(N * sizeof(int), device);
|
||||
int *d_B = (int *)omp_target_alloc(N * sizeof(int), device);
|
||||
int *d_C = (int *)omp_target_alloc(N * sizeof(int), device);
|
||||
|
||||
// CHECK: omp_target_alloc succeeded
|
||||
printf("omp_target_alloc %s\n", d_A && d_B && d_C ? "succeeded" : "failed");
|
||||
|
||||
omp_target_memcpy(d_B, pB, N * sizeof(int), 0, 0, device,
|
||||
omp_get_initial_device());
|
||||
omp_target_memcpy(d_C, pC, N * sizeof(int), 0, 0, device,
|
||||
omp_get_initial_device());
|
||||
|
||||
#pragma omp target is_device_ptr(d_A, d_B, d_C) device(device)
|
||||
{
|
||||
#pragma omp parallel for schedule(static, 1)
|
||||
for (int i = 0; i < N; i++) {
|
||||
d_A[i] = d_B[i] + d_C[i] + 1;
|
||||
}
|
||||
}
|
||||
|
||||
omp_target_memcpy(pA, d_A, N * sizeof(int), 0, 0, omp_get_initial_device(),
|
||||
device);
|
||||
|
||||
// CHECK: Test omp_target_memcpy: Succeeded
|
||||
int fail = 0;
|
||||
for (int i = 0; i < N; ++i) {
|
||||
if (A[i] != i + 2)
|
||||
fail++;
|
||||
}
|
||||
if (fail) {
|
||||
printf("Test omp_target_memcpy: Failed\n");
|
||||
} else {
|
||||
printf("Test omp_target_memcpy: Succeeded\n");
|
||||
}
|
||||
|
||||
//
|
||||
// target_is_present and target_associate/disassociate_ptr
|
||||
//
|
||||
init(A, B, C);
|
||||
|
||||
// CHECK: B is not present, associating it...
|
||||
// CHECK: omp_target_associate_ptr B succeeded
|
||||
if (!omp_target_is_present(B, device)) {
|
||||
printf("B is not present, associating it...\n");
|
||||
int rc = omp_target_associate_ptr(B, d_B, N * sizeof(int), 0, device);
|
||||
printf("omp_target_associate_ptr B %s\n", !rc ? "succeeded" : "failed");
|
||||
}
|
||||
|
||||
// CHECK: C is not present, associating it...
|
||||
// CHECK: omp_target_associate_ptr C succeeded
|
||||
if (!omp_target_is_present(C, device)) {
|
||||
printf("C is not present, associating it...\n");
|
||||
int rc = omp_target_associate_ptr(C, d_C, N * sizeof(int), 0, device);
|
||||
printf("omp_target_associate_ptr C %s\n", !rc ? "succeeded" : "failed");
|
||||
}
|
||||
|
||||
// CHECK: Inside target data: A is not present
|
||||
// CHECK: Inside target data: B is present
|
||||
// CHECK: Inside target data: C is present
|
||||
#pragma omp target data map(from : B, C) device(device)
|
||||
{
|
||||
printf("Inside target data: A is%s present\n",
|
||||
omp_target_is_present(A, device) ? "" : " not");
|
||||
printf("Inside target data: B is%s present\n",
|
||||
omp_target_is_present(B, device) ? "" : " not");
|
||||
printf("Inside target data: C is%s present\n",
|
||||
omp_target_is_present(C, device) ? "" : " not");
|
||||
|
||||
#pragma omp target map(from : A) device(device)
|
||||
{
|
||||
#pragma omp parallel for schedule(static, 1)
|
||||
for (int i = 0; i < N; i++)
|
||||
A[i] = B[i] + C[i] + 1;
|
||||
}
|
||||
}
|
||||
|
||||
// CHECK: B is present, disassociating it...
|
||||
// CHECK: omp_target_disassociate_ptr B succeeded
|
||||
// CHECK: C is present, disassociating it...
|
||||
// CHECK: omp_target_disassociate_ptr C succeeded
|
||||
if (omp_target_is_present(B, device)) {
|
||||
printf("B is present, disassociating it...\n");
|
||||
int rc = omp_target_disassociate_ptr(B, device);
|
||||
printf("omp_target_disassociate_ptr B %s\n", !rc ? "succeeded" : "failed");
|
||||
}
|
||||
if (omp_target_is_present(C, device)) {
|
||||
printf("C is present, disassociating it...\n");
|
||||
int rc = omp_target_disassociate_ptr(C, device);
|
||||
printf("omp_target_disassociate_ptr C %s\n", !rc ? "succeeded" : "failed");
|
||||
}
|
||||
|
||||
// CHECK: Test omp_target_associate_ptr: Succeeded
|
||||
fail = 0;
|
||||
for (int i = 0; i < N; ++i) {
|
||||
if (A[i] != i + 2)
|
||||
fail++;
|
||||
}
|
||||
if (fail) {
|
||||
printf("Test omp_target_associate_ptr: Failed\n");
|
||||
} else {
|
||||
printf("Test omp_target_associate_ptr: Succeeded\n");
|
||||
}
|
||||
|
||||
omp_target_free(d_A, device);
|
||||
omp_target_free(d_B, device);
|
||||
omp_target_free(d_C, device);
|
||||
|
||||
printf("Done!\n");
|
||||
|
||||
return 0;
|
||||
}
|
||||
@@ -1,95 +0,0 @@
|
||||
// RUN: %libomptarget-compile-run-and-check-aarch64-unknown-linux-gnu
|
||||
// RUN: %libomptarget-compile-run-and-check-powerpc64-ibm-linux-gnu
|
||||
// RUN: %libomptarget-compile-run-and-check-powerpc64le-ibm-linux-gnu
|
||||
// RUN: %libomptarget-compile-run-and-check-x86_64-pc-linux-gnu
|
||||
|
||||
// UNSUPPORTED: clang-6, clang-7, clang-8, clang-9
|
||||
|
||||
#include <omp.h>
|
||||
#include <stdio.h>
|
||||
|
||||
#pragma omp requires unified_shared_memory
|
||||
|
||||
#define N 1024
|
||||
|
||||
int main(int argc, char *argv[]) {
|
||||
int fails;
|
||||
void *host_alloc = 0, *device_alloc = 0;
|
||||
int *a = (int *)malloc(N * sizeof(int));
|
||||
|
||||
// Init
|
||||
for (int i = 0; i < N; ++i) {
|
||||
a[i] = 10;
|
||||
}
|
||||
host_alloc = &a[0];
|
||||
|
||||
//
|
||||
// map + target no close
|
||||
//
|
||||
#pragma omp target data map(tofrom : a[ : N]) map(tofrom : device_alloc)
|
||||
{
|
||||
#pragma omp target map(tofrom : device_alloc)
|
||||
{ device_alloc = &a[0]; }
|
||||
}
|
||||
|
||||
// CHECK: a used from unified memory.
|
||||
if (device_alloc == host_alloc)
|
||||
printf("a used from unified memory.\n");
|
||||
|
||||
//
|
||||
// map + target with close
|
||||
//
|
||||
device_alloc = 0;
|
||||
#pragma omp target data map(close, tofrom : a[ : N]) map(tofrom : device_alloc)
|
||||
{
|
||||
#pragma omp target map(tofrom : device_alloc)
|
||||
{ device_alloc = &a[0]; }
|
||||
}
|
||||
// CHECK: a copied to device.
|
||||
if (device_alloc != host_alloc)
|
||||
printf("a copied to device.\n");
|
||||
|
||||
//
|
||||
// map + use_device_ptr no close
|
||||
//
|
||||
device_alloc = 0;
|
||||
#pragma omp target data map(tofrom : a[ : N]) use_device_ptr(a)
|
||||
{ device_alloc = &a[0]; }
|
||||
|
||||
// CHECK: a used from unified memory with use_device_ptr.
|
||||
if (device_alloc == host_alloc)
|
||||
printf("a used from unified memory with use_device_ptr.\n");
|
||||
|
||||
//
|
||||
// map + use_device_ptr close
|
||||
//
|
||||
device_alloc = 0;
|
||||
#pragma omp target data map(close, tofrom : a[ : N]) use_device_ptr(a)
|
||||
{ device_alloc = &a[0]; }
|
||||
|
||||
// CHECK: a used from device memory with use_device_ptr.
|
||||
if (device_alloc != host_alloc)
|
||||
printf("a used from device memory with use_device_ptr.\n");
|
||||
|
||||
//
|
||||
// map enter/exit + close
|
||||
//
|
||||
device_alloc = 0;
|
||||
#pragma omp target enter data map(close, to : a[ : N])
|
||||
|
||||
#pragma omp target map(from : device_alloc)
|
||||
{ device_alloc = &a[0]; }
|
||||
|
||||
#pragma omp target exit data map(from : a[ : N])
|
||||
|
||||
// CHECK: a has been mapped to the device.
|
||||
if (device_alloc != host_alloc)
|
||||
printf("a has been mapped to the device.\n");
|
||||
|
||||
free(a);
|
||||
|
||||
// CHECK: Done!
|
||||
printf("Done!\n");
|
||||
|
||||
return 0;
|
||||
}
|
||||
@@ -1,86 +0,0 @@
|
||||
// RUN: %libomptarget-compile-run-and-check-aarch64-unknown-linux-gnu
|
||||
// RUN: %libomptarget-compile-run-and-check-powerpc64-ibm-linux-gnu
|
||||
// RUN: %libomptarget-compile-run-and-check-powerpc64le-ibm-linux-gnu
|
||||
// RUN: %libomptarget-compile-run-and-check-x86_64-pc-linux-gnu
|
||||
|
||||
#include <omp.h>
|
||||
#include <stdio.h>
|
||||
|
||||
// ---------------------------------------------------------------------------
|
||||
// Various definitions copied from OpenMP RTL
|
||||
|
||||
extern void __tgt_register_requires(int64_t);
|
||||
|
||||
extern void __tgt_target_data_begin(int64_t device_id, int32_t arg_num,
|
||||
void **args_base, void **args,
|
||||
int64_t *arg_sizes, int64_t *arg_types);
|
||||
|
||||
extern void __tgt_target_data_end(int64_t device_id, int32_t arg_num,
|
||||
void **args_base, void **args,
|
||||
int64_t *arg_sizes, int64_t *arg_types);
|
||||
|
||||
// End of definitions copied from OpenMP RTL.
|
||||
// ---------------------------------------------------------------------------
|
||||
|
||||
#pragma omp requires unified_shared_memory
|
||||
|
||||
#define N 1024
|
||||
|
||||
int main(int argc, char *argv[]) {
|
||||
int fails;
|
||||
void *host_alloc = 0, *device_alloc = 0;
|
||||
int *a = (int *)malloc(N * sizeof(int));
|
||||
|
||||
// Manual registration of requires flags for Clang versions
|
||||
// that do not support requires.
|
||||
__tgt_register_requires(8);
|
||||
|
||||
// Init
|
||||
for (int i = 0; i < N; ++i) {
|
||||
a[i] = 10;
|
||||
}
|
||||
host_alloc = &a[0];
|
||||
|
||||
// Dummy target region that ensures the runtime library is loaded when
|
||||
// the target data begin/end functions are manually called below.
|
||||
#pragma omp target
|
||||
{}
|
||||
|
||||
// Manual calls
|
||||
int device_id = omp_get_default_device();
|
||||
int arg_num = 1;
|
||||
void **args_base = (void **)&a;
|
||||
void **args = (void **)&a;
|
||||
int64_t arg_sizes[arg_num];
|
||||
|
||||
arg_sizes[0] = sizeof(int) * N;
|
||||
|
||||
int64_t arg_types[arg_num];
|
||||
|
||||
// Ox400 enables the CLOSE map type in the runtime:
|
||||
// OMP_TGT_MAPTYPE_CLOSE = 0x400
|
||||
// OMP_TGT_MAPTYPE_TO = 0x001
|
||||
arg_types[0] = 0x400 | 0x001;
|
||||
|
||||
device_alloc = host_alloc;
|
||||
|
||||
__tgt_target_data_begin(device_id, arg_num, args_base, args, arg_sizes,
|
||||
arg_types);
|
||||
|
||||
#pragma omp target data use_device_ptr(a)
|
||||
{ device_alloc = a; }
|
||||
|
||||
__tgt_target_data_end(device_id, arg_num, args_base, args, arg_sizes,
|
||||
arg_types);
|
||||
|
||||
// CHECK: a was copied to the device
|
||||
if (device_alloc != host_alloc)
|
||||
printf("a was copied to the device\n");
|
||||
|
||||
free(a);
|
||||
|
||||
// CHECK: Done!
|
||||
printf("Done!\n");
|
||||
|
||||
return 0;
|
||||
}
|
||||
@@ -1,135 +0,0 @@
|
||||
// RUN: %libomptarget-compile-run-and-check-aarch64-unknown-linux-gnu
|
||||
// RUN: %libomptarget-compile-run-and-check-powerpc64-ibm-linux-gnu
|
||||
// RUN: %libomptarget-compile-run-and-check-powerpc64le-ibm-linux-gnu
|
||||
// RUN: %libomptarget-compile-run-and-check-x86_64-pc-linux-gnu
|
||||
|
||||
// UNSUPPORTED: clang-6, clang-7, clang-8, clang-9
|
||||
|
||||
#include <omp.h>
|
||||
#include <stdio.h>
|
||||
|
||||
#pragma omp requires unified_shared_memory
|
||||
|
||||
#define N 1024
|
||||
|
||||
int main(int argc, char *argv[]) {
|
||||
int fails;
|
||||
void *host_alloc, *device_alloc;
|
||||
void *host_data, *device_data;
|
||||
int *alloc = (int *)malloc(N * sizeof(int));
|
||||
int data[N];
|
||||
|
||||
for (int i = 0; i < N; ++i) {
|
||||
alloc[i] = 10;
|
||||
data[i] = 1;
|
||||
}
|
||||
|
||||
host_data = &data[0];
|
||||
host_alloc = &alloc[0];
|
||||
|
||||
//
|
||||
// Test that updates on the device are not visible to host
|
||||
// when only a TO mapping is used.
|
||||
//
|
||||
#pragma omp target map(tofrom \
|
||||
: device_data, device_alloc) map(close, to \
|
||||
: alloc[:N], data \
|
||||
[:N])
|
||||
{
|
||||
device_data = &data[0];
|
||||
device_alloc = &alloc[0];
|
||||
|
||||
for (int i = 0; i < N; i++) {
|
||||
alloc[i] += 1;
|
||||
data[i] += 1;
|
||||
}
|
||||
}
|
||||
|
||||
// CHECK: Address of alloc on device different from host address.
|
||||
if (device_alloc != host_alloc)
|
||||
printf("Address of alloc on device different from host address.\n");
|
||||
|
||||
// CHECK: Address of data on device different from host address.
|
||||
if (device_data != host_data)
|
||||
printf("Address of data on device different from host address.\n");
|
||||
|
||||
// On the host, check that the arrays have been updated.
|
||||
// CHECK: Alloc host values not updated: Succeeded
|
||||
fails = 0;
|
||||
for (int i = 0; i < N; i++) {
|
||||
if (alloc[i] != 10)
|
||||
fails++;
|
||||
}
|
||||
printf("Alloc host values not updated: %s\n",
|
||||
(fails == 0) ? "Succeeded" : "Failed");
|
||||
|
||||
// CHECK: Data host values not updated: Succeeded
|
||||
fails = 0;
|
||||
for (int i = 0; i < N; i++) {
|
||||
if (data[i] != 1)
|
||||
fails++;
|
||||
}
|
||||
printf("Data host values not updated: %s\n",
|
||||
(fails == 0) ? "Succeeded" : "Failed");
|
||||
|
||||
//
|
||||
// Test that updates on the device are visible on host
|
||||
// when a from is used.
|
||||
//
|
||||
|
||||
for (int i = 0; i < N; i++) {
|
||||
alloc[i] += 1;
|
||||
data[i] += 1;
|
||||
}
|
||||
|
||||
#pragma omp target map(close, tofrom : alloc[:N], data[:N])
|
||||
{
|
||||
// CHECK: Alloc device values are correct: Succeeded
|
||||
fails = 0;
|
||||
for (int i = 0; i < N; i++) {
|
||||
if (alloc[i] != 11)
|
||||
fails++;
|
||||
}
|
||||
printf("Alloc device values are correct: %s\n",
|
||||
(fails == 0) ? "Succeeded" : "Failed");
|
||||
// CHECK: Data device values are correct: Succeeded
|
||||
fails = 0;
|
||||
for (int i = 0; i < N; i++) {
|
||||
if (data[i] != 2)
|
||||
fails++;
|
||||
}
|
||||
printf("Data device values are correct: %s\n",
|
||||
(fails == 0) ? "Succeeded" : "Failed");
|
||||
|
||||
// Update values on the device
|
||||
for (int i = 0; i < N; i++) {
|
||||
alloc[i] += 1;
|
||||
data[i] += 1;
|
||||
}
|
||||
}
|
||||
|
||||
// CHECK: Alloc host values updated: Succeeded
|
||||
fails = 0;
|
||||
for (int i = 0; i < N; i++) {
|
||||
if (alloc[i] != 12)
|
||||
fails++;
|
||||
}
|
||||
printf("Alloc host values updated: %s\n",
|
||||
(fails == 0) ? "Succeeded" : "Failed");
|
||||
|
||||
// CHECK: Data host values updated: Succeeded
|
||||
fails = 0;
|
||||
for (int i = 0; i < N; i++) {
|
||||
if (data[i] != 3)
|
||||
fails++;
|
||||
}
|
||||
printf("Data host values updated: %s\n",
|
||||
(fails == 0) ? "Succeeded" : "Failed");
|
||||
|
||||
free(alloc);
|
||||
|
||||
// CHECK: Done!
|
||||
printf("Done!\n");
|
||||
|
||||
return 0;
|
||||
}
|
||||
@@ -1,114 +0,0 @@
|
||||
// RUN: %libomptarget-compile-run-and-check-aarch64-unknown-linux-gnu
|
||||
// RUN: %libomptarget-compile-run-and-check-powerpc64-ibm-linux-gnu
|
||||
// RUN: %libomptarget-compile-run-and-check-powerpc64le-ibm-linux-gnu
|
||||
// RUN: %libomptarget-compile-run-and-check-x86_64-pc-linux-gnu
|
||||
|
||||
#include <stdio.h>
|
||||
#include <omp.h>
|
||||
|
||||
// ---------------------------------------------------------------------------
|
||||
// Various definitions copied from OpenMP RTL
|
||||
|
||||
extern void __tgt_register_requires(int64_t);
|
||||
|
||||
// End of definitions copied from OpenMP RTL.
|
||||
// ---------------------------------------------------------------------------
|
||||
|
||||
#pragma omp requires unified_shared_memory
|
||||
|
||||
#define N 1024
|
||||
|
||||
int main(int argc, char *argv[]) {
|
||||
int fails;
|
||||
void *host_alloc, *device_alloc;
|
||||
void *host_data, *device_data;
|
||||
int *alloc = (int *)malloc(N * sizeof(int));
|
||||
int data[N];
|
||||
|
||||
// Manual registration of requires flags for Clang versions
|
||||
// that do not support requires.
|
||||
__tgt_register_requires(8);
|
||||
|
||||
for (int i = 0; i < N; ++i) {
|
||||
alloc[i] = 10;
|
||||
data[i] = 1;
|
||||
}
|
||||
|
||||
host_data = &data[0];
|
||||
host_alloc = &alloc[0];
|
||||
|
||||
// implicit mapping of data
|
||||
#pragma omp target map(tofrom : device_data, device_alloc)
|
||||
{
|
||||
device_data = &data[0];
|
||||
device_alloc = &alloc[0];
|
||||
|
||||
for (int i = 0; i < N; i++) {
|
||||
alloc[i] += 1;
|
||||
data[i] += 1;
|
||||
}
|
||||
}
|
||||
|
||||
// CHECK: Address of alloc on device matches host address.
|
||||
if (device_alloc == host_alloc)
|
||||
printf("Address of alloc on device matches host address.\n");
|
||||
|
||||
// CHECK: Address of data on device matches host address.
|
||||
if (device_data == host_data)
|
||||
printf("Address of data on device matches host address.\n");
|
||||
|
||||
// On the host, check that the arrays have been updated.
|
||||
// CHECK: Alloc device values updated: Succeeded
|
||||
fails = 0;
|
||||
for (int i = 0; i < N; i++) {
|
||||
if (alloc[i] != 11)
|
||||
fails++;
|
||||
}
|
||||
printf("Alloc device values updated: %s\n",
|
||||
(fails == 0) ? "Succeeded" : "Failed");
|
||||
|
||||
// CHECK: Data device values updated: Succeeded
|
||||
fails = 0;
|
||||
for (int i = 0; i < N; i++) {
|
||||
if (data[i] != 2)
|
||||
fails++;
|
||||
}
|
||||
printf("Data device values updated: %s\n",
|
||||
(fails == 0) ? "Succeeded" : "Failed");
|
||||
|
||||
//
|
||||
// Test that updates on the host snd on the device are both visible.
|
||||
//
|
||||
|
||||
// Update on the host.
|
||||
for (int i = 0; i < N; ++i) {
|
||||
alloc[i] += 1;
|
||||
data[i] += 1;
|
||||
}
|
||||
|
||||
#pragma omp target
|
||||
{
|
||||
// CHECK: Alloc host values updated: Succeeded
|
||||
fails = 0;
|
||||
for (int i = 0; i < N; i++) {
|
||||
if (alloc[i] != 12)
|
||||
fails++;
|
||||
}
|
||||
printf("Alloc host values updated: %s\n",
|
||||
(fails == 0) ? "Succeeded" : "Failed");
|
||||
// CHECK: Data host values updated: Succeeded
|
||||
fails = 0;
|
||||
for (int i = 0; i < N; i++) {
|
||||
if (data[i] != 3)
|
||||
fails++;
|
||||
}
|
||||
printf("Data host values updated: %s\n",
|
||||
(fails == 0) ? "Succeeded" : "Failed");
|
||||
}
|
||||
|
||||
free(alloc);
|
||||
|
||||
printf("Done!\n");
|
||||
|
||||
return 0;
|
||||
}
|
||||
@@ -30,7 +30,7 @@ if(${OPENMP_STANDALONE_BUILD})
|
||||
# If adding a new architecture, take a look at cmake/LibompGetArchitecture.cmake
|
||||
libomp_get_architecture(LIBOMP_DETECTED_ARCH)
|
||||
set(LIBOMP_ARCH ${LIBOMP_DETECTED_ARCH} CACHE STRING
|
||||
"The architecture to build for (x86_64/i386/arm/ppc64/ppc64le/aarch64/mic/mips/mips64/riscv64).")
|
||||
"The architecture to build for (x86_64/i386/arm/ppc64/ppc64le/aarch64/mic/mips/mips64).")
|
||||
# Should assertions be enabled? They are on by default.
|
||||
set(LIBOMP_ENABLE_ASSERTIONS TRUE CACHE BOOL
|
||||
"enable assertions?")
|
||||
@@ -58,15 +58,13 @@ else() # Part of LLVM build
|
||||
set(LIBOMP_ARCH aarch64)
|
||||
elseif(LIBOMP_NATIVE_ARCH MATCHES "arm")
|
||||
set(LIBOMP_ARCH arm)
|
||||
elseif(LIBOMP_NATIVE_ARCH MATCHES "riscv64")
|
||||
set(LIBOMP_ARCH riscv64)
|
||||
else()
|
||||
# last ditch effort
|
||||
libomp_get_architecture(LIBOMP_ARCH)
|
||||
endif ()
|
||||
set(LIBOMP_ENABLE_ASSERTIONS ${LLVM_ENABLE_ASSERTIONS})
|
||||
endif()
|
||||
libomp_check_variable(LIBOMP_ARCH 32e x86_64 32 i386 arm ppc64 ppc64le aarch64 mic mips mips64 riscv64)
|
||||
libomp_check_variable(LIBOMP_ARCH 32e x86_64 32 i386 arm ppc64 ppc64le aarch64 mic mips mips64)
|
||||
|
||||
set(LIBOMP_LIB_TYPE normal CACHE STRING
|
||||
"Performance,Profiling,Stubs library (normal/profile/stubs)")
|
||||
@@ -97,6 +95,8 @@ if(NOT DEFINED CMAKE_MACOSX_RPATH)
|
||||
endif()
|
||||
|
||||
# User specified flags. These are appended to the configured flags.
|
||||
set(LIBOMP_CFLAGS "" CACHE STRING
|
||||
"Appended user specified C compiler flags.")
|
||||
set(LIBOMP_CXXFLAGS "" CACHE STRING
|
||||
"Appended user specified C++ compiler flags.")
|
||||
set(LIBOMP_CPPFLAGS "" CACHE STRING
|
||||
@@ -142,7 +142,6 @@ set(PPC64 FALSE)
|
||||
set(MIC FALSE)
|
||||
set(MIPS64 FALSE)
|
||||
set(MIPS FALSE)
|
||||
set(RISCV64 FALSE)
|
||||
if("${LIBOMP_ARCH}" STREQUAL "i386" OR "${LIBOMP_ARCH}" STREQUAL "32") # IA-32 architecture
|
||||
set(IA32 TRUE)
|
||||
elseif("${LIBOMP_ARCH}" STREQUAL "x86_64" OR "${LIBOMP_ARCH}" STREQUAL "32e") # Intel(R) 64 architecture
|
||||
@@ -163,8 +162,6 @@ elseif("${LIBOMP_ARCH}" STREQUAL "mips") # MIPS architecture
|
||||
set(MIPS TRUE)
|
||||
elseif("${LIBOMP_ARCH}" STREQUAL "mips64") # MIPS64 architecture
|
||||
set(MIPS64 TRUE)
|
||||
elseif("${LIBOMP_ARCH}" STREQUAL "riscv64") # RISCV64 architecture
|
||||
set(RISCV64 TRUE)
|
||||
endif()
|
||||
|
||||
# Set some flags based on build_type
|
||||
|
||||
@@ -53,7 +53,6 @@ Architectures Supported
|
||||
* IBM(R) Power architecture (big endian)
|
||||
* IBM(R) Power architecture (little endian)
|
||||
* MIPS and MIPS64 architecture
|
||||
* RISCV64 architecture
|
||||
|
||||
Supported RTL Build Configurations
|
||||
==================================
|
||||
|
||||
@@ -45,8 +45,6 @@ function(libomp_get_architecture return_arch)
|
||||
#error ARCHITECTURE=mips64
|
||||
#elif defined(__mips__) && !defined(__mips64)
|
||||
#error ARCHITECTURE=mips
|
||||
#elif defined(__riscv) && __riscv_xlen == 64
|
||||
#error ARCHITECTURE=riscv64
|
||||
#else
|
||||
#error ARCHITECTURE=UnknownArchitecture
|
||||
#endif
|
||||
|
||||
@@ -22,20 +22,30 @@ macro(libomp_setup_flags flags)
|
||||
endif()
|
||||
endmacro()
|
||||
|
||||
# C++ compiler flags
|
||||
function(libomp_get_cxxflags cxxflags)
|
||||
# Gets flags common to both the C and C++ compiler
|
||||
function(libomp_get_c_and_cxxflags_common flags)
|
||||
set(flags_local)
|
||||
libomp_append(flags_local -fno-exceptions LIBOMP_HAVE_FNO_EXCEPTIONS_FLAG)
|
||||
libomp_append(flags_local -fno-rtti LIBOMP_HAVE_FNO_RTTI_FLAG)
|
||||
libomp_append(flags_local -Wno-class-memaccess LIBOMP_HAVE_WNO_CLASS_MEMACCESS_FLAG)
|
||||
libomp_append(flags_local -Wno-covered-switch-default LIBOMP_HAVE_WNO_COVERED_SWITCH_DEFAULT_FLAG)
|
||||
libomp_append(flags_local -Wno-frame-address LIBOMP_HAVE_WNO_FRAME_ADDRESS_FLAG)
|
||||
libomp_append(flags_local -Wno-strict-aliasing LIBOMP_HAVE_WNO_STRICT_ALIASING_FLAG)
|
||||
libomp_append(flags_local -Wstringop-overflow=0 LIBOMP_HAVE_WSTRINGOP_OVERFLOW_FLAG)
|
||||
libomp_append(flags_local -Wno-stringop-truncation LIBOMP_HAVE_WNO_STRINGOP_TRUNCATION_FLAG)
|
||||
if(${OPENMP_STANDALONE_BUILD})
|
||||
libomp_append(flags_local -Wsign-compare LIBOMP_HAVE_WNO_SIGN_COMPARE_FLAG)
|
||||
libomp_append(flags_local -Wunused-function LIBOMP_HAVE_WNO_UNUSED_FUNCTION_FLAG)
|
||||
libomp_append(flags_local -Wunused-local-typedef LIBOMP_HAVE_WNO_UNUSED_LOCAL_TYPEDEF_FLAG)
|
||||
libomp_append(flags_local -Wunused-value LIBOMP_HAVE_WNO_UNUSED_VALUE_FLAG)
|
||||
libomp_append(flags_local -Wunused-variable LIBOMP_HAVE_WNO_UNUSED_VARIABLE_FLAG)
|
||||
libomp_append(flags_local -Wdeprecated-register LIBOMP_HAVE_WNO_DEPRECATED_REGISTER_FLAG)
|
||||
libomp_append(flags_local -Wunknown-pragmas LIBOMP_HAVE_WNO_UNKNOWN_PRAGMAS_FLAG)
|
||||
libomp_append(flags_local -Wcomment LIBOMP_HAVE_WNO_COMMENT_FLAG)
|
||||
libomp_append(flags_local -Wself-assign LIBOMP_HAVE_WNO_SELF_ASSIGN_FLAG)
|
||||
libomp_append(flags_local -Wformat-pedantic LIBOMP_HAVE_WNO_FORMAT_PEDANTIC_FLAG)
|
||||
endif()
|
||||
libomp_append(flags_local -Wno-switch LIBOMP_HAVE_WNO_SWITCH_FLAG)
|
||||
libomp_append(flags_local -Wno-uninitialized LIBOMP_HAVE_WNO_UNINITIALIZED_FLAG)
|
||||
libomp_append(flags_local -Wno-unused-but-set-variable LIBOMP_HAVE_WNO_UNUSED_BUT_SET_VARIABLE_FLAG)
|
||||
libomp_append(flags_local -Wno-covered-switch-default LIBOMP_HAVE_WNO_COVERED_SWITCH_DEFAULT_FLAG)
|
||||
libomp_append(flags_local -Wno-gnu-anonymous-struct LIBOMP_HAVE_WNO_GNU_ANONYMOUS_STRUCT_FLAG)
|
||||
libomp_append(flags_local -Wno-missing-field-initializers LIBOMP_HAVE_WNO_MISSING_FIELD_INITIALIZERS_FLAG)
|
||||
libomp_append(flags_local -Wno-missing-braces LIBOMP_HAVE_WNO_MISSING_BRACES_FLAG)
|
||||
libomp_append(flags_local -Wno-vla-extension LIBOMP_HAVE_WNO_VLA_EXTENSION_FLAG)
|
||||
libomp_append(flags_local -Wstringop-overflow=0 LIBOMP_HAVE_WSTRINGOP_OVERFLOW_FLAG)
|
||||
libomp_append(flags_local /GS LIBOMP_HAVE_GS_FLAG)
|
||||
libomp_append(flags_local /EHsc LIBOMP_HAVE_EHSC_FLAG)
|
||||
libomp_append(flags_local /Oy- LIBOMP_HAVE_OY__FLAG)
|
||||
@@ -61,7 +71,29 @@ function(libomp_get_cxxflags cxxflags)
|
||||
libomp_append(flags_local -ftls-model=initial-exec LIBOMP_HAVE_FTLS_MODEL_FLAG)
|
||||
libomp_append(flags_local "-opt-streaming-stores never" LIBOMP_HAVE_OPT_STREAMING_STORES_FLAG)
|
||||
endif()
|
||||
set(cxxflags_local ${flags_local} ${LIBOMP_CXXFLAGS})
|
||||
set(${flags} ${flags_local} PARENT_SCOPE)
|
||||
endfunction()
|
||||
|
||||
# C compiler flags
|
||||
function(libomp_get_cflags cflags)
|
||||
set(cflags_local)
|
||||
libomp_get_c_and_cxxflags_common(cflags_local)
|
||||
# flags only for the C Compiler
|
||||
libomp_append(cflags_local /TP LIBOMP_HAVE_TP_FLAG)
|
||||
libomp_append(cflags_local "-x c++" LIBOMP_HAVE_X_CPP_FLAG)
|
||||
set(cflags_local ${cflags_local} ${LIBOMP_CFLAGS})
|
||||
libomp_setup_flags(cflags_local)
|
||||
set(${cflags} ${cflags_local} PARENT_SCOPE)
|
||||
endfunction()
|
||||
|
||||
# C++ compiler flags
|
||||
function(libomp_get_cxxflags cxxflags)
|
||||
set(cxxflags_local)
|
||||
libomp_get_c_and_cxxflags_common(cxxflags_local)
|
||||
if(${OPENMP_STANDALONE_BUILD})
|
||||
libomp_append(cxxflags_local -Wcast-qual LIBOMP_HAVE_WCAST_QUAL_FLAG)
|
||||
endif()
|
||||
set(cxxflags_local ${cxxflags_local} ${LIBOMP_CXXFLAGS})
|
||||
libomp_setup_flags(cxxflags_local)
|
||||
set(${cxxflags} ${cxxflags_local} PARENT_SCOPE)
|
||||
endfunction()
|
||||
@@ -126,11 +158,11 @@ function(libomp_get_libflags libflags)
|
||||
if(${IA32})
|
||||
libomp_append(libflags_local -lirc_pic LIBOMP_HAVE_IRC_PIC_LIBRARY)
|
||||
endif()
|
||||
if(${CMAKE_SYSTEM_NAME} MATCHES "DragonFly|FreeBSD")
|
||||
if(${CMAKE_SYSTEM_NAME} MATCHES "DragonFly")
|
||||
libomp_append(libflags_local "-Wl,--no-as-needed" LIBOMP_HAVE_AS_NEEDED_FLAG)
|
||||
libomp_append(libflags_local "-lm")
|
||||
libomp_append(libflags_local "-Wl,--as-needed" LIBOMP_HAVE_AS_NEEDED_FLAG)
|
||||
elseif(${CMAKE_SYSTEM_NAME} MATCHES "NetBSD")
|
||||
elseif(${CMAKE_SYSTEM_NAME} MATCHES "(Free|Net)BSD")
|
||||
libomp_append(libflags_local -lm)
|
||||
endif()
|
||||
set(libflags_local ${libflags_local} ${LIBOMP_LIBFLAGS})
|
||||
|
||||
@@ -211,9 +211,6 @@ else()
|
||||
elseif(${MIPS} OR ${MIPS64})
|
||||
libomp_append(libomp_expected_library_deps libc.so.6)
|
||||
libomp_append(libomp_expected_library_deps ld.so.1)
|
||||
elseif(${RISCV64})
|
||||
libomp_append(libomp_expected_library_deps libc.so.6)
|
||||
libomp_append(libomp_expected_library_deps ld.so.1)
|
||||
endif()
|
||||
libomp_append(libomp_expected_library_deps libpthread.so.0 IF_FALSE STUBS_LIBRARY)
|
||||
libomp_append(libomp_expected_library_deps libhwloc.so.5 LIBOMP_USE_HWLOC)
|
||||
|
||||
@@ -105,8 +105,6 @@ function(libomp_get_legal_arch return_arch_string)
|
||||
set(${return_arch_string} "MIPS" PARENT_SCOPE)
|
||||
elseif(${MIPS64})
|
||||
set(${return_arch_string} "MIPS64" PARENT_SCOPE)
|
||||
elseif(${RISCV64})
|
||||
set(${return_arch_string} "RISCV64" PARENT_SCOPE)
|
||||
else()
|
||||
set(${return_arch_string} "${LIBOMP_ARCH}" PARENT_SCOPE)
|
||||
libomp_warning_say("libomp_get_legal_arch(): Warning: Unknown architecture: Using ${LIBOMP_ARCH}")
|
||||
|
||||
@@ -45,32 +45,43 @@ function(libomp_check_architecture_flag flag retval)
|
||||
set(${retval} ${${retval}} PARENT_SCOPE)
|
||||
endfunction()
|
||||
|
||||
# Checking CXX, Linker Flags
|
||||
# Checking C, CXX, Linker Flags
|
||||
check_cxx_compiler_flag(-fno-exceptions LIBOMP_HAVE_FNO_EXCEPTIONS_FLAG)
|
||||
check_cxx_compiler_flag(-fno-rtti LIBOMP_HAVE_FNO_RTTI_FLAG)
|
||||
check_cxx_compiler_flag(-Wno-class-memaccess LIBOMP_HAVE_WNO_CLASS_MEMACCESS_FLAG)
|
||||
check_cxx_compiler_flag(-Wno-covered-switch-default LIBOMP_HAVE_WNO_COVERED_SWITCH_DEFAULT_FLAG)
|
||||
check_cxx_compiler_flag(-Wno-frame-address LIBOMP_HAVE_WNO_FRAME_ADDRESS_FLAG)
|
||||
check_cxx_compiler_flag(-Wno-strict-aliasing LIBOMP_HAVE_WNO_STRICT_ALIASING_FLAG)
|
||||
check_cxx_compiler_flag(-Wstringop-overflow=0 LIBOMP_HAVE_WSTRINGOP_OVERFLOW_FLAG)
|
||||
check_cxx_compiler_flag(-Wno-stringop-truncation LIBOMP_HAVE_WNO_STRINGOP_TRUNCATION_FLAG)
|
||||
check_cxx_compiler_flag(-Wno-switch LIBOMP_HAVE_WNO_SWITCH_FLAG)
|
||||
check_cxx_compiler_flag(-Wno-uninitialized LIBOMP_HAVE_WNO_UNINITIALIZED_FLAG)
|
||||
check_cxx_compiler_flag(-Wno-unused-but-set-variable LIBOMP_HAVE_WNO_UNUSED_BUT_SET_VARIABLE_FLAG)
|
||||
check_cxx_compiler_flag(-msse2 LIBOMP_HAVE_MSSE2_FLAG)
|
||||
check_cxx_compiler_flag(-ftls-model=initial-exec LIBOMP_HAVE_FTLS_MODEL_FLAG)
|
||||
check_c_compiler_flag("-x c++" LIBOMP_HAVE_X_CPP_FLAG)
|
||||
check_cxx_compiler_flag(-Wcast-qual LIBOMP_HAVE_WCAST_QUAL_FLAG)
|
||||
check_c_compiler_flag(-Wunused-function LIBOMP_HAVE_WNO_UNUSED_FUNCTION_FLAG)
|
||||
check_c_compiler_flag(-Wunused-local-typedef LIBOMP_HAVE_WNO_UNUSED_LOCAL_TYPEDEF_FLAG)
|
||||
check_c_compiler_flag(-Wunused-value LIBOMP_HAVE_WNO_UNUSED_VALUE_FLAG)
|
||||
check_c_compiler_flag(-Wunused-variable LIBOMP_HAVE_WNO_UNUSED_VARIABLE_FLAG)
|
||||
check_c_compiler_flag(-Wswitch LIBOMP_HAVE_WNO_SWITCH_FLAG)
|
||||
check_c_compiler_flag(-Wcovered-switch-default LIBOMP_HAVE_WNO_COVERED_SWITCH_DEFAULT_FLAG)
|
||||
check_c_compiler_flag(-Wdeprecated-register LIBOMP_HAVE_WNO_DEPRECATED_REGISTER_FLAG)
|
||||
check_c_compiler_flag(-Wsign-compare LIBOMP_HAVE_WNO_SIGN_COMPARE_FLAG)
|
||||
check_c_compiler_flag(-Wgnu-anonymous-struct LIBOMP_HAVE_WNO_GNU_ANONYMOUS_STRUCT_FLAG)
|
||||
check_c_compiler_flag(-Wunknown-pragmas LIBOMP_HAVE_WNO_UNKNOWN_PRAGMAS_FLAG)
|
||||
check_c_compiler_flag(-Wmissing-field-initializers LIBOMP_HAVE_WNO_MISSING_FIELD_INITIALIZERS_FLAG)
|
||||
check_c_compiler_flag(-Wmissing-braces LIBOMP_HAVE_WNO_MISSING_BRACES_FLAG)
|
||||
check_c_compiler_flag(-Wcomment LIBOMP_HAVE_WNO_COMMENT_FLAG)
|
||||
check_c_compiler_flag(-Wself-assign LIBOMP_HAVE_WNO_SELF_ASSIGN_FLAG)
|
||||
check_c_compiler_flag(-Wvla-extension LIBOMP_HAVE_WNO_VLA_EXTENSION_FLAG)
|
||||
check_c_compiler_flag(-Wformat-pedantic LIBOMP_HAVE_WNO_FORMAT_PEDANTIC_FLAG)
|
||||
check_c_compiler_flag(-Wstringop-overflow=0 LIBOMP_HAVE_WSTRINGOP_OVERFLOW_FLAG)
|
||||
check_c_compiler_flag(-msse2 LIBOMP_HAVE_MSSE2_FLAG)
|
||||
check_c_compiler_flag(-ftls-model=initial-exec LIBOMP_HAVE_FTLS_MODEL_FLAG)
|
||||
libomp_check_architecture_flag(-mmic LIBOMP_HAVE_MMIC_FLAG)
|
||||
libomp_check_architecture_flag(-m32 LIBOMP_HAVE_M32_FLAG)
|
||||
if(WIN32)
|
||||
if(MSVC)
|
||||
# Check Windows MSVC style flags.
|
||||
check_c_compiler_flag(/TP LIBOMP_HAVE_TP_FLAG)
|
||||
check_cxx_compiler_flag(/EHsc LIBOMP_HAVE_EHSC_FLAG)
|
||||
check_cxx_compiler_flag(/GS LIBOMP_HAVE_GS_FLAG)
|
||||
check_cxx_compiler_flag(/Oy- LIBOMP_HAVE_Oy__FLAG)
|
||||
check_cxx_compiler_flag(/arch:SSE2 LIBOMP_HAVE_ARCH_SSE2_FLAG)
|
||||
check_cxx_compiler_flag(/Qsafeseh LIBOMP_HAVE_QSAFESEH_FLAG)
|
||||
endif()
|
||||
check_cxx_compiler_flag(-mrtm LIBOMP_HAVE_MRTM_FLAG)
|
||||
check_c_compiler_flag(-mrtm LIBOMP_HAVE_MRTM_FLAG)
|
||||
# It is difficult to create a dummy masm assembly file
|
||||
# and then check the MASM assembler to see if these flags exist and work,
|
||||
# so we assume they do for Windows.
|
||||
@@ -235,8 +246,7 @@ else()
|
||||
# (LIBOMP_ARCH STREQUAL arm) OR
|
||||
(LIBOMP_ARCH STREQUAL aarch64) OR
|
||||
(LIBOMP_ARCH STREQUAL ppc64le) OR
|
||||
(LIBOMP_ARCH STREQUAL ppc64) OR
|
||||
(LIBOMP_ARCH STREQUAL riscv64))
|
||||
(LIBOMP_ARCH STREQUAL ppc64))
|
||||
AND # OS supported?
|
||||
((WIN32 AND LIBOMP_HAVE_PSAPI) OR APPLE OR (NOT WIN32 AND LIBOMP_HAVE_WEAK_ATTRIBUTE)))
|
||||
set(LIBOMP_HAVE_OMPT_SUPPORT TRUE)
|
||||
|
||||
@@ -31,7 +31,7 @@ add_custom_command(
|
||||
|
||||
# Set the -D definitions for all sources
|
||||
# UNICODE and _UNICODE are set in LLVM's CMake system. They affect the
|
||||
# ittnotify code and should only be set when compiling ittnotify_static.cpp
|
||||
# ittnotify code and should only be set when compiling ittnotify_static.c
|
||||
# on Windows (done below).
|
||||
# TODO: Fix the UNICODE usage in ittnotify code for Windows.
|
||||
remove_definitions(-DUNICODE -D_UNICODE)
|
||||
@@ -51,10 +51,11 @@ if(${LIBOMP_USE_HWLOC})
|
||||
endif()
|
||||
|
||||
# Getting correct source files to build library
|
||||
set(LIBOMP_CFILES)
|
||||
set(LIBOMP_CXXFILES)
|
||||
set(LIBOMP_ASMFILES)
|
||||
if(STUBS_LIBRARY)
|
||||
set(LIBOMP_CXXFILES kmp_stub.cpp)
|
||||
if(${STUBS_LIBRARY})
|
||||
set(LIBOMP_CFILES kmp_stub.cpp)
|
||||
else()
|
||||
# Get C++ files
|
||||
set(LIBOMP_CXXFILES
|
||||
@@ -92,7 +93,7 @@ else()
|
||||
libomp_append(LIBOMP_CXXFILES kmp_gsupport.cpp)
|
||||
libomp_append(LIBOMP_ASMFILES z_Linux_asm.S) # Unix assembly file
|
||||
endif()
|
||||
libomp_append(LIBOMP_CXXFILES thirdparty/ittnotify/ittnotify_static.cpp LIBOMP_USE_ITT_NOTIFY)
|
||||
libomp_append(LIBOMP_CFILES thirdparty/ittnotify/ittnotify_static.c LIBOMP_USE_ITT_NOTIFY)
|
||||
libomp_append(LIBOMP_CXXFILES kmp_debugger.cpp LIBOMP_USE_DEBUGGER)
|
||||
libomp_append(LIBOMP_CXXFILES kmp_stats.cpp LIBOMP_STATS)
|
||||
libomp_append(LIBOMP_CXXFILES kmp_stats_timing.cpp LIBOMP_STATS)
|
||||
@@ -106,14 +107,16 @@ libomp_append(LIBOMP_CXXFILES kmp_version.cpp)
|
||||
libomp_append(LIBOMP_CXXFILES ompt-general.cpp IF_TRUE LIBOMP_OMPT_SUPPORT)
|
||||
libomp_append(LIBOMP_CXXFILES tsan_annotations.cpp IF_TRUE LIBOMP_TSAN_SUPPORT)
|
||||
|
||||
set(LIBOMP_SOURCE_FILES ${LIBOMP_CXXFILES} ${LIBOMP_ASMFILES})
|
||||
set(LIBOMP_SOURCE_FILES ${LIBOMP_CFILES} ${LIBOMP_CXXFILES} ${LIBOMP_ASMFILES})
|
||||
# For Windows, there is a resource file (.rc -> .res) that is also compiled
|
||||
libomp_append(LIBOMP_SOURCE_FILES libomp.rc WIN32)
|
||||
|
||||
# Get compiler and assembler flags
|
||||
libomp_get_cflags(LIBOMP_CONFIGURED_CFLAGS)
|
||||
libomp_get_cxxflags(LIBOMP_CONFIGURED_CXXFLAGS)
|
||||
libomp_get_asmflags(LIBOMP_CONFIGURED_ASMFLAGS)
|
||||
# Set the compiler flags for each type of source
|
||||
set_source_files_properties(${LIBOMP_CFILES} PROPERTIES COMPILE_FLAGS "${LIBOMP_CONFIGURED_CFLAGS}")
|
||||
set_source_files_properties(${LIBOMP_CXXFILES} PROPERTIES COMPILE_FLAGS "${LIBOMP_CONFIGURED_CXXFLAGS}")
|
||||
set_source_files_properties(${LIBOMP_ASMFILES} PROPERTIES COMPILE_FLAGS "${LIBOMP_CONFIGURED_ASMFLAGS}")
|
||||
# Let the compiler handle the assembly files on Unix-like systems
|
||||
@@ -188,12 +191,12 @@ if(WIN32)
|
||||
libomp_append(LIBOMP_MASM_DEFINITIONS "-DOMPT_SUPPORT" IF_TRUE_1_0 LIBOMP_OMPT_SUPPORT)
|
||||
libomp_list_to_string("${LIBOMP_MASM_DEFINITIONS}" LIBOMP_MASM_DEFINITIONS)
|
||||
set_property(SOURCE z_Windows_NT-586_asm.asm APPEND_STRING PROPERTY COMPILE_FLAGS " ${LIBOMP_MASM_DEFINITIONS}")
|
||||
set_source_files_properties(thirdparty/ittnotify/ittnotify_static.cpp PROPERTIES COMPILE_DEFINITIONS "UNICODE")
|
||||
set_source_files_properties(thirdparty/ittnotify/ittnotify_static.c PROPERTIES COMPILE_DEFINITIONS "UNICODE")
|
||||
|
||||
# Create Windows import library
|
||||
# the import library is "re-linked" to include kmp_import.cpp which prevents
|
||||
# linking of both Visual Studio OpenMP and newly built OpenMP
|
||||
set_source_files_properties(kmp_import.cpp PROPERTIES COMPILE_FLAGS "${LIBOMP_CONFIGURED_CXXFLAGS}")
|
||||
set_source_files_properties(kmp_import.cpp PROPERTIES COMPILE_FLAGS "${LIBOMP_CONFIGURED_CFLAGS}")
|
||||
set(LIBOMP_IMP_LIB_FILE ${LIBOMP_LIB_NAME}${CMAKE_STATIC_LIBRARY_SUFFIX})
|
||||
set(LIBOMP_GENERATED_IMP_LIB_FILENAME ${LIBOMP_LIB_FILE}${CMAKE_STATIC_LIBRARY_SUFFIX})
|
||||
set_target_properties(omp PROPERTIES
|
||||
|
||||
+3
-2
@@ -2181,9 +2181,10 @@ struct kmp_dephash_entry {
|
||||
typedef struct kmp_dephash {
|
||||
kmp_dephash_entry_t **buckets;
|
||||
size_t size;
|
||||
size_t generation;
|
||||
#ifdef KMP_DEBUG
|
||||
kmp_uint32 nelements;
|
||||
kmp_uint32 nconflicts;
|
||||
#endif
|
||||
} kmp_dephash_t;
|
||||
|
||||
typedef struct kmp_task_affinity_info {
|
||||
@@ -3341,7 +3342,7 @@ extern int __kmp_aux_set_affinity_mask_proc(int proc, void **mask);
|
||||
extern int __kmp_aux_unset_affinity_mask_proc(int proc, void **mask);
|
||||
extern int __kmp_aux_get_affinity_mask_proc(int proc, void **mask);
|
||||
extern void __kmp_balanced_affinity(kmp_info_t *th, int team_size);
|
||||
#if KMP_OS_LINUX || KMP_OS_FREEBSD
|
||||
#if KMP_OS_LINUX
|
||||
extern int kmp_set_thread_affinity_mask_initial(void);
|
||||
#endif
|
||||
#endif /* KMP_AFFINITY_SUPPORTED */
|
||||
|
||||
@@ -1968,7 +1968,7 @@ static void __kmp_dispatch_set_hierarchy_values() {
|
||||
__kmp_hier_max_units[kmp_hier_layer_e::LAYER_THREAD + 1] =
|
||||
nPackages * nCoresPerPkg * __kmp_nThreadsPerCore;
|
||||
__kmp_hier_max_units[kmp_hier_layer_e::LAYER_L1 + 1] = __kmp_ncores;
|
||||
#if KMP_ARCH_X86_64 && (KMP_OS_LINUX || KMP_OS_FREEBSD || KMP_OS_WINDOWS)
|
||||
#if KMP_ARCH_X86_64 && (KMP_OS_LINUX || KMP_OS_WINDOWS)
|
||||
if (__kmp_mic_type >= mic3)
|
||||
__kmp_hier_max_units[kmp_hier_layer_e::LAYER_L2 + 1] = __kmp_ncores / 2;
|
||||
else
|
||||
@@ -1982,7 +1982,7 @@ static void __kmp_dispatch_set_hierarchy_values() {
|
||||
__kmp_hier_threads_per[kmp_hier_layer_e::LAYER_THREAD + 1] = 1;
|
||||
__kmp_hier_threads_per[kmp_hier_layer_e::LAYER_L1 + 1] =
|
||||
__kmp_nThreadsPerCore;
|
||||
#if KMP_ARCH_X86_64 && (KMP_OS_LINUX || KMP_OS_FREEBSD || KMP_OS_WINDOWS)
|
||||
#if KMP_ARCH_X86_64 && (KMP_OS_LINUX || KMP_OS_WINDOWS)
|
||||
if (__kmp_mic_type >= mic3)
|
||||
__kmp_hier_threads_per[kmp_hier_layer_e::LAYER_L2 + 1] =
|
||||
2 * __kmp_nThreadsPerCore;
|
||||
|
||||
@@ -160,7 +160,6 @@ public:
|
||||
};
|
||||
#endif /* KMP_USE_HWLOC */
|
||||
|
||||
#if KMP_OS_LINUX || KMP_OS_FREEBSD
|
||||
#if KMP_OS_LINUX
|
||||
/* On some of the older OS's that we build on, these constants aren't present
|
||||
in <asm/unistd.h> #included from <sys.syscall.h>. They must be the same on
|
||||
@@ -235,10 +234,6 @@ public:
|
||||
#endif /* __NR_sched_getaffinity */
|
||||
#error Unknown or unsupported architecture
|
||||
#endif /* KMP_ARCH_* */
|
||||
#elif KMP_OS_FREEBSD
|
||||
#include <pthread.h>
|
||||
#include <pthread_np.h>
|
||||
#endif
|
||||
class KMPNativeAffinity : public KMPAffinity {
|
||||
class Mask : public KMPAffinity::Mask {
|
||||
typedef unsigned char mask_t;
|
||||
@@ -299,13 +294,8 @@ class KMPNativeAffinity : public KMPAffinity {
|
||||
int get_system_affinity(bool abort_on_error) override {
|
||||
KMP_ASSERT2(KMP_AFFINITY_CAPABLE(),
|
||||
"Illegal get affinity operation when not capable");
|
||||
#if KMP_OS_LINUX
|
||||
int retval =
|
||||
syscall(__NR_sched_getaffinity, 0, __kmp_affin_mask_size, mask);
|
||||
#elif KMP_OS_FREEBSD
|
||||
int retval =
|
||||
pthread_getaffinity_np(pthread_self(), __kmp_affin_mask_size, reinterpret_cast<cpuset_t *>(mask));
|
||||
#endif
|
||||
if (retval >= 0) {
|
||||
return 0;
|
||||
}
|
||||
@@ -318,13 +308,8 @@ class KMPNativeAffinity : public KMPAffinity {
|
||||
int set_system_affinity(bool abort_on_error) const override {
|
||||
KMP_ASSERT2(KMP_AFFINITY_CAPABLE(),
|
||||
"Illegal get affinity operation when not capable");
|
||||
#if KMP_OS_LINUX
|
||||
int retval =
|
||||
syscall(__NR_sched_setaffinity, 0, __kmp_affin_mask_size, mask);
|
||||
#elif KMP_OS_FREEBSD
|
||||
int retval =
|
||||
pthread_setaffinity_np(pthread_self(), __kmp_affin_mask_size, reinterpret_cast<cpuset_t *>(mask));
|
||||
#endif
|
||||
if (retval >= 0) {
|
||||
return 0;
|
||||
}
|
||||
@@ -362,7 +347,7 @@ class KMPNativeAffinity : public KMPAffinity {
|
||||
}
|
||||
api_type get_api_type() const override { return NATIVE_OS; }
|
||||
};
|
||||
#endif /* KMP_OS_LINUX || KMP_OS_FREEBSD */
|
||||
#endif /* KMP_OS_LINUX */
|
||||
|
||||
#if KMP_OS_WINDOWS
|
||||
class KMPNativeAffinity : public KMPAffinity {
|
||||
|
||||
@@ -545,8 +545,7 @@ void __kmpc_end_serialized_parallel(ident_t *loc, kmp_int32 global_tid) {
|
||||
if (ompt_enabled.ompt_callback_parallel_end) {
|
||||
ompt_callbacks.ompt_callback(ompt_callback_parallel_end)(
|
||||
&(serial_team->t.ompt_team_info.parallel_data), parent_task_data,
|
||||
ompt_parallel_invoker_program | ompt_parallel_team,
|
||||
OMPT_LOAD_RETURN_ADDRESS(global_tid));
|
||||
ompt_parallel_invoker_program, OMPT_LOAD_RETURN_ADDRESS(global_tid));
|
||||
}
|
||||
__ompt_lw_taskteam_unlink(this_thr);
|
||||
this_thr->th.ompt_thread_info.state = ompt_state_overhead;
|
||||
@@ -677,8 +676,7 @@ void __kmpc_flush(ident_t *loc) {
|
||||
#endif // KMP_COMPILER_ICC
|
||||
}
|
||||
#endif // KMP_MIC
|
||||
#elif (KMP_ARCH_ARM || KMP_ARCH_AARCH64 || KMP_ARCH_MIPS || KMP_ARCH_MIPS64 || \
|
||||
KMP_ARCH_RISCV64)
|
||||
#elif (KMP_ARCH_ARM || KMP_ARCH_AARCH64 || KMP_ARCH_MIPS || KMP_ARCH_MIPS64)
|
||||
// Nothing to see here move along
|
||||
#elif KMP_ARCH_PPC64
|
||||
// Nothing needed here (we have a real MB above).
|
||||
|
||||
@@ -633,25 +633,5 @@
|
||||
GOMP_loop_ull_doacross_guided_start
|
||||
#define KMP_API_NAME_GOMP_LOOP_ULL_DOACROSS_RUNTIME_START \
|
||||
GOMP_loop_ull_doacross_runtime_start
|
||||
#define KMP_API_NAME_GOMP_LOOP_NONMONOTONIC_DYNAMIC_NEXT \
|
||||
GOMP_loop_nonmonotonic_dynamic_next
|
||||
#define KMP_API_NAME_GOMP_LOOP_NONMONOTONIC_DYNAMIC_START \
|
||||
GOMP_loop_nonmonotonic_dynamic_start
|
||||
#define KMP_API_NAME_GOMP_LOOP_NONMONOTONIC_GUIDED_NEXT \
|
||||
GOMP_loop_nonmonotonic_guided_next
|
||||
#define KMP_API_NAME_GOMP_LOOP_NONMONOTONIC_GUIDED_START \
|
||||
GOMP_loop_nonmonotonic_guided_start
|
||||
#define KMP_API_NAME_GOMP_LOOP_ULL_NONMONOTONIC_DYNAMIC_NEXT \
|
||||
GOMP_loop_ull_nonmonotonic_dynamic_next
|
||||
#define KMP_API_NAME_GOMP_LOOP_ULL_NONMONOTONIC_DYNAMIC_START \
|
||||
GOMP_loop_ull_nonmonotonic_dynamic_start
|
||||
#define KMP_API_NAME_GOMP_LOOP_ULL_NONMONOTONIC_GUIDED_NEXT \
|
||||
GOMP_loop_ull_nonmonotonic_guided_next
|
||||
#define KMP_API_NAME_GOMP_LOOP_ULL_NONMONOTONIC_GUIDED_START \
|
||||
GOMP_loop_ull_nonmonotonic_guided_start
|
||||
#define KMP_API_NAME_GOMP_PARALLEL_LOOP_NONMONOTONIC_DYNAMIC \
|
||||
GOMP_parallel_loop_nonmonotonic_dynamic
|
||||
#define KMP_API_NAME_GOMP_PARALLEL_LOOP_NONMONOTONIC_GUIDED \
|
||||
GOMP_parallel_loop_nonmonotonic_guided
|
||||
|
||||
#endif /* KMP_FTN_OS_H */
|
||||
|
||||
@@ -431,7 +431,7 @@ std::atomic<int> __kmp_thread_pool_active_nth = ATOMIC_VAR_INIT(0);
|
||||
/* -------------------------------------------------
|
||||
* GLOBAL/ROOT STATE */
|
||||
KMP_ALIGN_CACHE
|
||||
kmp_global_t __kmp_global;
|
||||
kmp_global_t __kmp_global = {{0}};
|
||||
|
||||
/* ----------------------------------------------- */
|
||||
/* GLOBAL SYNCHRONIZATION LOCKS */
|
||||
|
||||
@@ -22,7 +22,7 @@ extern "C" {
|
||||
#endif // __cplusplus
|
||||
|
||||
#define MKLOC(loc, routine) \
|
||||
static ident_t loc = {0, KMP_IDENT_KMPC, 0, 0, ";unknown;unknown;0;0;;"};
|
||||
static ident_t(loc) = {0, KMP_IDENT_KMPC, 0, 0, ";unknown;unknown;0;0;;"};
|
||||
|
||||
#include "kmp_ftn_os.h"
|
||||
|
||||
@@ -622,16 +622,10 @@ LOOP_START(KMP_EXPAND_NAME(KMP_API_NAME_GOMP_LOOP_STATIC_START), kmp_sch_static)
|
||||
LOOP_NEXT(KMP_EXPAND_NAME(KMP_API_NAME_GOMP_LOOP_STATIC_NEXT), {})
|
||||
LOOP_START(KMP_EXPAND_NAME(KMP_API_NAME_GOMP_LOOP_DYNAMIC_START),
|
||||
kmp_sch_dynamic_chunked)
|
||||
LOOP_START(KMP_EXPAND_NAME(KMP_API_NAME_GOMP_LOOP_NONMONOTONIC_DYNAMIC_START),
|
||||
kmp_sch_dynamic_chunked)
|
||||
LOOP_NEXT(KMP_EXPAND_NAME(KMP_API_NAME_GOMP_LOOP_DYNAMIC_NEXT), {})
|
||||
LOOP_NEXT(KMP_EXPAND_NAME(KMP_API_NAME_GOMP_LOOP_NONMONOTONIC_DYNAMIC_NEXT), {})
|
||||
LOOP_START(KMP_EXPAND_NAME(KMP_API_NAME_GOMP_LOOP_GUIDED_START),
|
||||
kmp_sch_guided_chunked)
|
||||
LOOP_START(KMP_EXPAND_NAME(KMP_API_NAME_GOMP_LOOP_NONMONOTONIC_GUIDED_START),
|
||||
kmp_sch_guided_chunked)
|
||||
LOOP_NEXT(KMP_EXPAND_NAME(KMP_API_NAME_GOMP_LOOP_GUIDED_NEXT), {})
|
||||
LOOP_NEXT(KMP_EXPAND_NAME(KMP_API_NAME_GOMP_LOOP_NONMONOTONIC_GUIDED_NEXT), {})
|
||||
LOOP_RUNTIME_START(KMP_EXPAND_NAME(KMP_API_NAME_GOMP_LOOP_RUNTIME_START),
|
||||
kmp_sch_runtime)
|
||||
LOOP_NEXT(KMP_EXPAND_NAME(KMP_API_NAME_GOMP_LOOP_RUNTIME_NEXT), {})
|
||||
@@ -898,16 +892,6 @@ LOOP_NEXT_ULL(KMP_EXPAND_NAME(KMP_API_NAME_GOMP_LOOP_ULL_DYNAMIC_NEXT), {})
|
||||
LOOP_START_ULL(KMP_EXPAND_NAME(KMP_API_NAME_GOMP_LOOP_ULL_GUIDED_START),
|
||||
kmp_sch_guided_chunked)
|
||||
LOOP_NEXT_ULL(KMP_EXPAND_NAME(KMP_API_NAME_GOMP_LOOP_ULL_GUIDED_NEXT), {})
|
||||
LOOP_START_ULL(
|
||||
KMP_EXPAND_NAME(KMP_API_NAME_GOMP_LOOP_ULL_NONMONOTONIC_DYNAMIC_START),
|
||||
kmp_sch_dynamic_chunked)
|
||||
LOOP_NEXT_ULL(
|
||||
KMP_EXPAND_NAME(KMP_API_NAME_GOMP_LOOP_ULL_NONMONOTONIC_DYNAMIC_NEXT), {})
|
||||
LOOP_START_ULL(
|
||||
KMP_EXPAND_NAME(KMP_API_NAME_GOMP_LOOP_ULL_NONMONOTONIC_GUIDED_START),
|
||||
kmp_sch_guided_chunked)
|
||||
LOOP_NEXT_ULL(
|
||||
KMP_EXPAND_NAME(KMP_API_NAME_GOMP_LOOP_ULL_NONMONOTONIC_GUIDED_NEXT), {})
|
||||
LOOP_RUNTIME_START_ULL(
|
||||
KMP_EXPAND_NAME(KMP_API_NAME_GOMP_LOOP_ULL_RUNTIME_START), kmp_sch_runtime)
|
||||
LOOP_NEXT_ULL(KMP_EXPAND_NAME(KMP_API_NAME_GOMP_LOOP_ULL_RUNTIME_NEXT), {})
|
||||
@@ -1503,12 +1487,6 @@ PARALLEL_LOOP(KMP_EXPAND_NAME(KMP_API_NAME_GOMP_PARALLEL_LOOP_STATIC),
|
||||
kmp_sch_static, OMPT_LOOP_PRE, OMPT_LOOP_POST)
|
||||
PARALLEL_LOOP(KMP_EXPAND_NAME(KMP_API_NAME_GOMP_PARALLEL_LOOP_DYNAMIC),
|
||||
kmp_sch_dynamic_chunked, OMPT_LOOP_PRE, OMPT_LOOP_POST)
|
||||
PARALLEL_LOOP(
|
||||
KMP_EXPAND_NAME(KMP_API_NAME_GOMP_PARALLEL_LOOP_NONMONOTONIC_GUIDED),
|
||||
kmp_sch_guided_chunked, OMPT_LOOP_PRE, OMPT_LOOP_POST)
|
||||
PARALLEL_LOOP(
|
||||
KMP_EXPAND_NAME(KMP_API_NAME_GOMP_PARALLEL_LOOP_NONMONOTONIC_DYNAMIC),
|
||||
kmp_sch_dynamic_chunked, OMPT_LOOP_PRE, OMPT_LOOP_POST)
|
||||
PARALLEL_LOOP(KMP_EXPAND_NAME(KMP_API_NAME_GOMP_PARALLEL_LOOP_GUIDED),
|
||||
kmp_sch_guided_chunked, OMPT_LOOP_PRE, OMPT_LOOP_POST)
|
||||
PARALLEL_LOOP(KMP_EXPAND_NAME(KMP_API_NAME_GOMP_PARALLEL_LOOP_RUNTIME),
|
||||
@@ -1964,26 +1942,6 @@ KMP_VERSION_SYMBOL(KMP_API_NAME_GOMP_LOOP_ULL_DOACROSS_GUIDED_START, 45,
|
||||
"GOMP_4.5");
|
||||
KMP_VERSION_SYMBOL(KMP_API_NAME_GOMP_LOOP_ULL_DOACROSS_RUNTIME_START, 45,
|
||||
"GOMP_4.5");
|
||||
KMP_VERSION_SYMBOL(KMP_API_NAME_GOMP_LOOP_NONMONOTONIC_DYNAMIC_START, 45,
|
||||
"GOMP_4.5");
|
||||
KMP_VERSION_SYMBOL(KMP_API_NAME_GOMP_LOOP_NONMONOTONIC_DYNAMIC_NEXT, 45,
|
||||
"GOMP_4.5");
|
||||
KMP_VERSION_SYMBOL(KMP_API_NAME_GOMP_LOOP_NONMONOTONIC_GUIDED_START, 45,
|
||||
"GOMP_4.5");
|
||||
KMP_VERSION_SYMBOL(KMP_API_NAME_GOMP_LOOP_NONMONOTONIC_GUIDED_NEXT, 45,
|
||||
"GOMP_4.5");
|
||||
KMP_VERSION_SYMBOL(KMP_API_NAME_GOMP_LOOP_ULL_NONMONOTONIC_DYNAMIC_START, 45,
|
||||
"GOMP_4.5");
|
||||
KMP_VERSION_SYMBOL(KMP_API_NAME_GOMP_LOOP_ULL_NONMONOTONIC_DYNAMIC_NEXT, 45,
|
||||
"GOMP_4.5");
|
||||
KMP_VERSION_SYMBOL(KMP_API_NAME_GOMP_LOOP_ULL_NONMONOTONIC_GUIDED_START, 45,
|
||||
"GOMP_4.5");
|
||||
KMP_VERSION_SYMBOL(KMP_API_NAME_GOMP_LOOP_ULL_NONMONOTONIC_GUIDED_NEXT, 45,
|
||||
"GOMP_4.5");
|
||||
KMP_VERSION_SYMBOL(KMP_API_NAME_GOMP_PARALLEL_LOOP_NONMONOTONIC_DYNAMIC, 45,
|
||||
"GOMP_4.5");
|
||||
KMP_VERSION_SYMBOL(KMP_API_NAME_GOMP_PARALLEL_LOOP_NONMONOTONIC_GUIDED, 45,
|
||||
"GOMP_4.5");
|
||||
|
||||
#endif // KMP_USE_VERSION_SYMBOLS
|
||||
|
||||
|
||||
@@ -2943,10 +2943,10 @@ static int (*direct_test_check[])(kmp_dyna_lock_t *, kmp_int32) = {
|
||||
#undef expand
|
||||
|
||||
// Exposes only one set of jump tables (*lock or *lock_with_checks).
|
||||
void (**__kmp_direct_destroy)(kmp_dyna_lock_t *) = 0;
|
||||
int (**__kmp_direct_set)(kmp_dyna_lock_t *, kmp_int32) = 0;
|
||||
int (**__kmp_direct_unset)(kmp_dyna_lock_t *, kmp_int32) = 0;
|
||||
int (**__kmp_direct_test)(kmp_dyna_lock_t *, kmp_int32) = 0;
|
||||
void (*(*__kmp_direct_destroy))(kmp_dyna_lock_t *) = 0;
|
||||
int (*(*__kmp_direct_set))(kmp_dyna_lock_t *, kmp_int32) = 0;
|
||||
int (*(*__kmp_direct_unset))(kmp_dyna_lock_t *, kmp_int32) = 0;
|
||||
int (*(*__kmp_direct_test))(kmp_dyna_lock_t *, kmp_int32) = 0;
|
||||
|
||||
// Jump tables for the indirect lock functions
|
||||
#define expand(l, op) (void (*)(kmp_user_lock_p)) __kmp_##op##_##l##_##lock,
|
||||
@@ -2993,10 +2993,10 @@ static int (*indirect_test_check[])(kmp_user_lock_p, kmp_int32) = {
|
||||
#undef expand
|
||||
|
||||
// Exposes only one jump tables (*lock or *lock_with_checks).
|
||||
void (**__kmp_indirect_destroy)(kmp_user_lock_p) = 0;
|
||||
int (**__kmp_indirect_set)(kmp_user_lock_p, kmp_int32) = 0;
|
||||
int (**__kmp_indirect_unset)(kmp_user_lock_p, kmp_int32) = 0;
|
||||
int (**__kmp_indirect_test)(kmp_user_lock_p, kmp_int32) = 0;
|
||||
void (*(*__kmp_indirect_destroy))(kmp_user_lock_p) = 0;
|
||||
int (*(*__kmp_indirect_set))(kmp_user_lock_p, kmp_int32) = 0;
|
||||
int (*(*__kmp_indirect_unset))(kmp_user_lock_p, kmp_int32) = 0;
|
||||
int (*(*__kmp_indirect_test))(kmp_user_lock_p, kmp_int32) = 0;
|
||||
|
||||
// Lock index table.
|
||||
kmp_indirect_lock_table_t __kmp_i_lock_table;
|
||||
|
||||
@@ -1122,18 +1122,18 @@ typedef struct {
|
||||
// Function tables for direct locks. Set/unset/test differentiate functions
|
||||
// with/without consistency checking.
|
||||
extern void (*__kmp_direct_init[])(kmp_dyna_lock_t *, kmp_dyna_lockseq_t);
|
||||
extern void (**__kmp_direct_destroy)(kmp_dyna_lock_t *);
|
||||
extern int (**__kmp_direct_set)(kmp_dyna_lock_t *, kmp_int32);
|
||||
extern int (**__kmp_direct_unset)(kmp_dyna_lock_t *, kmp_int32);
|
||||
extern int (**__kmp_direct_test)(kmp_dyna_lock_t *, kmp_int32);
|
||||
extern void (*(*__kmp_direct_destroy))(kmp_dyna_lock_t *);
|
||||
extern int (*(*__kmp_direct_set))(kmp_dyna_lock_t *, kmp_int32);
|
||||
extern int (*(*__kmp_direct_unset))(kmp_dyna_lock_t *, kmp_int32);
|
||||
extern int (*(*__kmp_direct_test))(kmp_dyna_lock_t *, kmp_int32);
|
||||
|
||||
// Function tables for indirect locks. Set/unset/test differentiate functions
|
||||
// with/withuot consistency checking.
|
||||
extern void (*__kmp_indirect_init[])(kmp_user_lock_p);
|
||||
extern void (**__kmp_indirect_destroy)(kmp_user_lock_p);
|
||||
extern int (**__kmp_indirect_set)(kmp_user_lock_p, kmp_int32);
|
||||
extern int (**__kmp_indirect_unset)(kmp_user_lock_p, kmp_int32);
|
||||
extern int (**__kmp_indirect_test)(kmp_user_lock_p, kmp_int32);
|
||||
extern void (*(*__kmp_indirect_destroy))(kmp_user_lock_p);
|
||||
extern int (*(*__kmp_indirect_set))(kmp_user_lock_p, kmp_int32);
|
||||
extern int (*(*__kmp_indirect_unset))(kmp_user_lock_p, kmp_int32);
|
||||
extern int (*(*__kmp_indirect_test))(kmp_user_lock_p, kmp_int32);
|
||||
|
||||
// Extracts direct lock tag from a user lock pointer
|
||||
#define KMP_EXTRACT_D_TAG(l) \
|
||||
|
||||
@@ -69,7 +69,7 @@
|
||||
#error Unknown compiler
|
||||
#endif
|
||||
|
||||
#if (KMP_OS_LINUX || KMP_OS_WINDOWS || KMP_OS_FREEBSD) && !KMP_OS_CNK
|
||||
#if (KMP_OS_LINUX || KMP_OS_WINDOWS) && !KMP_OS_CNK
|
||||
#define KMP_AFFINITY_SUPPORTED 1
|
||||
#if KMP_OS_WINDOWS && KMP_ARCH_X86_64
|
||||
#define KMP_GROUP_AFFINITY 1
|
||||
@@ -165,8 +165,7 @@ typedef unsigned long long kmp_uint64;
|
||||
|
||||
#if KMP_ARCH_X86 || KMP_ARCH_ARM || KMP_ARCH_MIPS
|
||||
#define KMP_SIZE_T_SPEC KMP_UINT32_SPEC
|
||||
#elif KMP_ARCH_X86_64 || KMP_ARCH_PPC64 || KMP_ARCH_AARCH64 || \
|
||||
KMP_ARCH_MIPS64 || KMP_ARCH_RISCV64
|
||||
#elif KMP_ARCH_X86_64 || KMP_ARCH_PPC64 || KMP_ARCH_AARCH64 || KMP_ARCH_MIPS64
|
||||
#define KMP_SIZE_T_SPEC KMP_UINT64_SPEC
|
||||
#else
|
||||
#error "Can't determine size_t printf format specifier."
|
||||
@@ -841,7 +840,7 @@ extern kmp_real64 __kmp_xchg_real64(volatile kmp_real64 *p, kmp_real64 v);
|
||||
#endif /* KMP_OS_WINDOWS */
|
||||
|
||||
#if KMP_ARCH_PPC64 || KMP_ARCH_ARM || KMP_ARCH_AARCH64 || KMP_ARCH_MIPS || \
|
||||
KMP_ARCH_MIPS64 || KMP_ARCH_RISCV64
|
||||
KMP_ARCH_MIPS64
|
||||
#define KMP_MB() __sync_synchronize()
|
||||
#endif
|
||||
|
||||
|
||||
@@ -98,7 +98,6 @@
|
||||
#define KMP_ARCH_PPC64 (KMP_ARCH_PPC64_LE || KMP_ARCH_PPC64_BE)
|
||||
#define KMP_ARCH_MIPS 0
|
||||
#define KMP_ARCH_MIPS64 0
|
||||
#define KMP_ARCH_RISCV64 0
|
||||
|
||||
#if KMP_OS_WINDOWS
|
||||
#if defined(_M_AMD64) || defined(__x86_64)
|
||||
@@ -136,9 +135,6 @@
|
||||
#undef KMP_ARCH_MIPS
|
||||
#define KMP_ARCH_MIPS 1
|
||||
#endif
|
||||
#elif defined __riscv && __riscv_xlen == 64
|
||||
#undef KMP_ARCH_RISCV64
|
||||
#define KMP_ARCH_RISCV64 1
|
||||
#endif
|
||||
#endif
|
||||
|
||||
@@ -203,7 +199,7 @@
|
||||
// TODO: Fixme - This is clever, but really fugly
|
||||
#if (1 != \
|
||||
KMP_ARCH_X86 + KMP_ARCH_X86_64 + KMP_ARCH_ARM + KMP_ARCH_PPC64 + \
|
||||
KMP_ARCH_AARCH64 + KMP_ARCH_MIPS + KMP_ARCH_MIPS64 + KMP_ARCH_RISCV64)
|
||||
KMP_ARCH_AARCH64 + KMP_ARCH_MIPS + KMP_ARCH_MIPS64)
|
||||
#error Unknown or unsupported architecture
|
||||
#endif
|
||||
|
||||
|
||||
+74
-162
@@ -1190,8 +1190,8 @@ void __kmp_serialized_parallel(ident_t *loc, kmp_int32 global_tid) {
|
||||
|
||||
ompt_callbacks.ompt_callback(ompt_callback_parallel_begin)(
|
||||
&(parent_task_info->task_data), &(parent_task_info->frame),
|
||||
&ompt_parallel_data, team_size,
|
||||
ompt_parallel_invoker_program | ompt_parallel_team, codeptr);
|
||||
&ompt_parallel_data, team_size, ompt_parallel_invoker_program,
|
||||
codeptr);
|
||||
}
|
||||
}
|
||||
#endif // OMPT_SUPPORT
|
||||
@@ -1481,13 +1481,9 @@ int __kmp_fork_call(ident_t *loc, int gtid,
|
||||
int team_size = master_set_numthreads
|
||||
? master_set_numthreads
|
||||
: get__nproc_2(parent_team, master_tid);
|
||||
int flags = OMPT_INVOKER(call_context) |
|
||||
((microtask == (microtask_t)__kmp_teams_master)
|
||||
? ompt_parallel_league
|
||||
: ompt_parallel_team);
|
||||
ompt_callbacks.ompt_callback(ompt_callback_parallel_begin)(
|
||||
parent_task_data, ompt_frame, &ompt_parallel_data, team_size, flags,
|
||||
return_address);
|
||||
parent_task_data, ompt_frame, &ompt_parallel_data, team_size,
|
||||
OMPT_INVOKER(call_context), return_address);
|
||||
}
|
||||
master_th->th.ompt_thread_info.state = ompt_state_overhead;
|
||||
}
|
||||
@@ -1516,17 +1512,19 @@ int __kmp_fork_call(ident_t *loc, int gtid,
|
||||
// AC: we are in serialized parallel
|
||||
__kmpc_serialized_parallel(loc, gtid);
|
||||
KMP_DEBUG_ASSERT(parent_team->t.t_serialized > 1);
|
||||
|
||||
// AC: need this in order enquiry functions work
|
||||
// correctly, will restore at join time
|
||||
parent_team->t.t_serialized--;
|
||||
#if OMPT_SUPPORT
|
||||
void *dummy;
|
||||
void **exit_frame_p;
|
||||
void **exit_runtime_p;
|
||||
|
||||
ompt_lw_taskteam_t lw_taskteam;
|
||||
|
||||
if (ompt_enabled.enabled) {
|
||||
__ompt_lw_taskteam_init(&lw_taskteam, master_th, gtid,
|
||||
&ompt_parallel_data, return_address);
|
||||
exit_frame_p = &(lw_taskteam.ompt_task_info.frame.exit_frame.ptr);
|
||||
exit_runtime_p = &(lw_taskteam.ompt_task_info.frame.exit_frame.ptr);
|
||||
|
||||
__ompt_lw_taskteam_link(&lw_taskteam, master_th, 0);
|
||||
// don't use lw_taskteam after linking. content was swaped
|
||||
@@ -1534,23 +1532,19 @@ int __kmp_fork_call(ident_t *loc, int gtid,
|
||||
/* OMPT implicit task begin */
|
||||
implicit_task_data = OMPT_CUR_TASK_DATA(master_th);
|
||||
if (ompt_enabled.ompt_callback_implicit_task) {
|
||||
OMPT_CUR_TASK_INFO(master_th)
|
||||
->thread_num = __kmp_tid_from_gtid(gtid);
|
||||
ompt_callbacks.ompt_callback(ompt_callback_implicit_task)(
|
||||
ompt_scope_begin, OMPT_CUR_TEAM_DATA(master_th),
|
||||
implicit_task_data, 1,
|
||||
OMPT_CUR_TASK_INFO(master_th)->thread_num, ompt_task_implicit);
|
||||
implicit_task_data, 1, __kmp_tid_from_gtid(gtid), ompt_task_implicit); // TODO: Can this be ompt_task_initial?
|
||||
OMPT_CUR_TASK_INFO(master_th)
|
||||
->thread_num = __kmp_tid_from_gtid(gtid);
|
||||
}
|
||||
|
||||
/* OMPT state */
|
||||
master_th->th.ompt_thread_info.state = ompt_state_work_parallel;
|
||||
} else {
|
||||
exit_frame_p = &dummy;
|
||||
exit_runtime_p = &dummy;
|
||||
}
|
||||
#endif
|
||||
// AC: need to decrement t_serialized for enquiry functions to work
|
||||
// correctly, will restore at join time
|
||||
parent_team->t.t_serialized--;
|
||||
|
||||
{
|
||||
KMP_TIME_PARTITIONED_BLOCK(OMP_parallel);
|
||||
@@ -1558,27 +1552,26 @@ int __kmp_fork_call(ident_t *loc, int gtid,
|
||||
__kmp_invoke_microtask(microtask, gtid, 0, argc, parent_team->t.t_argv
|
||||
#if OMPT_SUPPORT
|
||||
,
|
||||
exit_frame_p
|
||||
exit_runtime_p
|
||||
#endif
|
||||
);
|
||||
}
|
||||
|
||||
#if OMPT_SUPPORT
|
||||
*exit_runtime_p = NULL;
|
||||
if (ompt_enabled.enabled) {
|
||||
*exit_frame_p = NULL;
|
||||
OMPT_CUR_TASK_INFO(master_th)->frame.exit_frame = ompt_data_none;
|
||||
if (ompt_enabled.ompt_callback_implicit_task) {
|
||||
ompt_callbacks.ompt_callback(ompt_callback_implicit_task)(
|
||||
ompt_scope_end, NULL, implicit_task_data, 1,
|
||||
OMPT_CUR_TASK_INFO(master_th)->thread_num, ompt_task_implicit);
|
||||
OMPT_CUR_TASK_INFO(master_th)->thread_num, ompt_task_implicit); // TODO: Can this be ompt_task_initial?
|
||||
}
|
||||
ompt_parallel_data = *OMPT_CUR_TEAM_DATA(master_th);
|
||||
__ompt_lw_taskteam_unlink(master_th);
|
||||
|
||||
if (ompt_enabled.ompt_callback_parallel_end) {
|
||||
ompt_callbacks.ompt_callback(ompt_callback_parallel_end)(
|
||||
&ompt_parallel_data, OMPT_CUR_TASK_DATA(master_th),
|
||||
OMPT_INVOKER(call_context) | ompt_parallel_team,
|
||||
return_address);
|
||||
OMPT_CUR_TEAM_DATA(master_th), OMPT_CUR_TASK_DATA(master_th),
|
||||
OMPT_INVOKER(call_context), return_address);
|
||||
}
|
||||
master_th->th.ompt_thread_info.state = ompt_state_overhead;
|
||||
}
|
||||
@@ -1593,15 +1586,6 @@ int __kmp_fork_call(ident_t *loc, int gtid,
|
||||
parent_team->t.t_level++;
|
||||
parent_team->t.t_def_allocator = master_th->th.th_def_allocator; // save
|
||||
|
||||
#if OMPT_SUPPORT
|
||||
if (ompt_enabled.enabled) {
|
||||
ompt_lw_taskteam_t lw_taskteam;
|
||||
__ompt_lw_taskteam_init(&lw_taskteam, master_th, gtid,
|
||||
&ompt_parallel_data, return_address);
|
||||
__ompt_lw_taskteam_link(&lw_taskteam, master_th, 1, true);
|
||||
}
|
||||
#endif
|
||||
|
||||
/* Change number of threads in the team if requested */
|
||||
if (master_set_numthreads) { // The parallel has num_threads clause
|
||||
if (master_set_numthreads < master_th->th.th_teams_size.nth) {
|
||||
@@ -1730,7 +1714,7 @@ int __kmp_fork_call(ident_t *loc, int gtid,
|
||||
|
||||
#if OMPT_SUPPORT
|
||||
void *dummy;
|
||||
void **exit_frame_p;
|
||||
void **exit_runtime_p;
|
||||
ompt_task_info_t *task_info;
|
||||
|
||||
ompt_lw_taskteam_t lw_taskteam;
|
||||
@@ -1743,21 +1727,19 @@ int __kmp_fork_call(ident_t *loc, int gtid,
|
||||
// don't use lw_taskteam after linking. content was swaped
|
||||
|
||||
task_info = OMPT_CUR_TASK_INFO(master_th);
|
||||
exit_frame_p = &(task_info->frame.exit_frame.ptr);
|
||||
exit_runtime_p = &(task_info->frame.exit_frame.ptr);
|
||||
if (ompt_enabled.ompt_callback_implicit_task) {
|
||||
OMPT_CUR_TASK_INFO(master_th)
|
||||
->thread_num = __kmp_tid_from_gtid(gtid);
|
||||
ompt_callbacks.ompt_callback(ompt_callback_implicit_task)(
|
||||
ompt_scope_begin, OMPT_CUR_TEAM_DATA(master_th),
|
||||
&(task_info->task_data), 1,
|
||||
OMPT_CUR_TASK_INFO(master_th)->thread_num,
|
||||
ompt_task_implicit);
|
||||
&(task_info->task_data), 1, __kmp_tid_from_gtid(gtid), ompt_task_implicit); // TODO: Can this be ompt_task_initial?
|
||||
OMPT_CUR_TASK_INFO(master_th)
|
||||
->thread_num = __kmp_tid_from_gtid(gtid);
|
||||
}
|
||||
|
||||
/* OMPT state */
|
||||
master_th->th.ompt_thread_info.state = ompt_state_work_parallel;
|
||||
} else {
|
||||
exit_frame_p = &dummy;
|
||||
exit_runtime_p = &dummy;
|
||||
}
|
||||
#endif
|
||||
|
||||
@@ -1768,27 +1750,25 @@ int __kmp_fork_call(ident_t *loc, int gtid,
|
||||
parent_team->t.t_argv
|
||||
#if OMPT_SUPPORT
|
||||
,
|
||||
exit_frame_p
|
||||
exit_runtime_p
|
||||
#endif
|
||||
);
|
||||
}
|
||||
|
||||
#if OMPT_SUPPORT
|
||||
if (ompt_enabled.enabled) {
|
||||
*exit_frame_p = NULL;
|
||||
exit_runtime_p = NULL;
|
||||
if (ompt_enabled.ompt_callback_implicit_task) {
|
||||
ompt_callbacks.ompt_callback(ompt_callback_implicit_task)(
|
||||
ompt_scope_end, NULL, &(task_info->task_data), 1,
|
||||
OMPT_CUR_TASK_INFO(master_th)->thread_num,
|
||||
ompt_task_implicit);
|
||||
OMPT_CUR_TASK_INFO(master_th)->thread_num, ompt_task_implicit); // TODO: Can this be ompt_task_initial?
|
||||
}
|
||||
ompt_parallel_data = *OMPT_CUR_TEAM_DATA(master_th);
|
||||
|
||||
__ompt_lw_taskteam_unlink(master_th);
|
||||
if (ompt_enabled.ompt_callback_parallel_end) {
|
||||
ompt_callbacks.ompt_callback(ompt_callback_parallel_end)(
|
||||
&ompt_parallel_data, parent_task_data,
|
||||
OMPT_INVOKER(call_context) | ompt_parallel_team,
|
||||
return_address);
|
||||
OMPT_CUR_TEAM_DATA(master_th), parent_task_data,
|
||||
OMPT_INVOKER(call_context), return_address);
|
||||
}
|
||||
master_th->th.ompt_thread_info.state = ompt_state_overhead;
|
||||
}
|
||||
@@ -1820,23 +1800,6 @@ int __kmp_fork_call(ident_t *loc, int gtid,
|
||||
team->t.t_level--;
|
||||
// AC: call special invoker for outer "parallel" of teams construct
|
||||
invoker(gtid);
|
||||
#if OMPT_SUPPORT
|
||||
if (ompt_enabled.enabled) {
|
||||
ompt_task_info_t *task_info = OMPT_CUR_TASK_INFO(master_th);
|
||||
if (ompt_enabled.ompt_callback_implicit_task) {
|
||||
ompt_callbacks.ompt_callback(ompt_callback_implicit_task)(
|
||||
ompt_scope_end, NULL, &(task_info->task_data), 0,
|
||||
OMPT_CUR_TASK_INFO(master_th)->thread_num, ompt_task_initial);
|
||||
}
|
||||
if (ompt_enabled.ompt_callback_parallel_end) {
|
||||
ompt_callbacks.ompt_callback(ompt_callback_parallel_end)(
|
||||
&ompt_parallel_data, parent_task_data,
|
||||
OMPT_INVOKER(call_context) | ompt_parallel_league,
|
||||
return_address);
|
||||
}
|
||||
master_th->th.ompt_thread_info.state = ompt_state_overhead;
|
||||
}
|
||||
#endif
|
||||
} else {
|
||||
argv = args;
|
||||
for (i = argc - 1; i >= 0; --i)
|
||||
@@ -1850,7 +1813,7 @@ int __kmp_fork_call(ident_t *loc, int gtid,
|
||||
|
||||
#if OMPT_SUPPORT
|
||||
void *dummy;
|
||||
void **exit_frame_p;
|
||||
void **exit_runtime_p;
|
||||
ompt_task_info_t *task_info;
|
||||
|
||||
ompt_lw_taskteam_t lw_taskteam;
|
||||
@@ -1861,15 +1824,14 @@ int __kmp_fork_call(ident_t *loc, int gtid,
|
||||
__ompt_lw_taskteam_link(&lw_taskteam, master_th, 0);
|
||||
// don't use lw_taskteam after linking. content was swaped
|
||||
task_info = OMPT_CUR_TASK_INFO(master_th);
|
||||
exit_frame_p = &(task_info->frame.exit_frame.ptr);
|
||||
exit_runtime_p = &(task_info->frame.exit_frame.ptr);
|
||||
|
||||
/* OMPT implicit task begin */
|
||||
implicit_task_data = OMPT_CUR_TASK_DATA(master_th);
|
||||
if (ompt_enabled.ompt_callback_implicit_task) {
|
||||
ompt_callbacks.ompt_callback(ompt_callback_implicit_task)(
|
||||
ompt_scope_begin, OMPT_CUR_TEAM_DATA(master_th),
|
||||
implicit_task_data, 1, __kmp_tid_from_gtid(gtid),
|
||||
ompt_task_implicit);
|
||||
implicit_task_data, 1, __kmp_tid_from_gtid(gtid), ompt_task_implicit); // TODO: Can this be ompt_task_initial?
|
||||
OMPT_CUR_TASK_INFO(master_th)
|
||||
->thread_num = __kmp_tid_from_gtid(gtid);
|
||||
}
|
||||
@@ -1877,7 +1839,7 @@ int __kmp_fork_call(ident_t *loc, int gtid,
|
||||
/* OMPT state */
|
||||
master_th->th.ompt_thread_info.state = ompt_state_work_parallel;
|
||||
} else {
|
||||
exit_frame_p = &dummy;
|
||||
exit_runtime_p = &dummy;
|
||||
}
|
||||
#endif
|
||||
|
||||
@@ -1887,19 +1849,18 @@ int __kmp_fork_call(ident_t *loc, int gtid,
|
||||
__kmp_invoke_microtask(microtask, gtid, 0, argc, args
|
||||
#if OMPT_SUPPORT
|
||||
,
|
||||
exit_frame_p
|
||||
exit_runtime_p
|
||||
#endif
|
||||
);
|
||||
}
|
||||
|
||||
#if OMPT_SUPPORT
|
||||
if (ompt_enabled.enabled) {
|
||||
*exit_frame_p = NULL;
|
||||
*exit_runtime_p = NULL;
|
||||
if (ompt_enabled.ompt_callback_implicit_task) {
|
||||
ompt_callbacks.ompt_callback(ompt_callback_implicit_task)(
|
||||
ompt_scope_end, NULL, &(task_info->task_data), 1,
|
||||
OMPT_CUR_TASK_INFO(master_th)->thread_num,
|
||||
ompt_task_implicit);
|
||||
OMPT_CUR_TASK_INFO(master_th)->thread_num, ompt_task_implicit); // TODO: Can this be ompt_task_initial?
|
||||
}
|
||||
|
||||
ompt_parallel_data = *OMPT_CUR_TEAM_DATA(master_th);
|
||||
@@ -1907,8 +1868,7 @@ int __kmp_fork_call(ident_t *loc, int gtid,
|
||||
if (ompt_enabled.ompt_callback_parallel_end) {
|
||||
ompt_callbacks.ompt_callback(ompt_callback_parallel_end)(
|
||||
&ompt_parallel_data, parent_task_data,
|
||||
OMPT_INVOKER(call_context) | ompt_parallel_team,
|
||||
return_address);
|
||||
OMPT_INVOKER(call_context), return_address);
|
||||
}
|
||||
master_th->th.ompt_thread_info.state = ompt_state_overhead;
|
||||
}
|
||||
@@ -2265,11 +2225,12 @@ static inline void __kmp_join_restore_state(kmp_info_t *thread,
|
||||
|
||||
static inline void __kmp_join_ompt(int gtid, kmp_info_t *thread,
|
||||
kmp_team_t *team, ompt_data_t *parallel_data,
|
||||
int flags, void *codeptr) {
|
||||
fork_context_e fork_context, void *codeptr) {
|
||||
ompt_task_info_t *task_info = __ompt_get_task_info_object(0);
|
||||
if (ompt_enabled.ompt_callback_parallel_end) {
|
||||
ompt_callbacks.ompt_callback(ompt_callback_parallel_end)(
|
||||
parallel_data, &(task_info->task_data), flags, codeptr);
|
||||
parallel_data, &(task_info->task_data), OMPT_INVOKER(fork_context),
|
||||
codeptr);
|
||||
}
|
||||
|
||||
task_info->frame.enter_frame = ompt_data_none;
|
||||
@@ -2302,7 +2263,6 @@ void __kmp_join_call(ident_t *loc, int gtid
|
||||
master_th->th.th_ident = loc;
|
||||
|
||||
#if OMPT_SUPPORT
|
||||
void *team_microtask = (void *)team->t.t_pkfn;
|
||||
if (ompt_enabled.enabled) {
|
||||
master_th->th.ompt_thread_info.state = ompt_state_overhead;
|
||||
}
|
||||
@@ -2392,25 +2352,10 @@ void __kmp_join_call(ident_t *loc, int gtid
|
||||
if (master_th->th.th_teams_microtask && !exit_teams &&
|
||||
team->t.t_pkfn != (microtask_t)__kmp_teams_master &&
|
||||
team->t.t_level == master_th->th.th_teams_level + 1) {
|
||||
// AC: We need to leave the team structure intact at the end of parallel
|
||||
// inside the teams construct, so that at the next parallel same (hot) team
|
||||
// works, only adjust nesting levels
|
||||
#if OMPT_SUPPORT
|
||||
ompt_data_t ompt_parallel_data = ompt_data_none;
|
||||
if (ompt_enabled.enabled) {
|
||||
ompt_task_info_t *task_info = __ompt_get_task_info_object(0);
|
||||
if (ompt_enabled.ompt_callback_implicit_task) {
|
||||
int ompt_team_size = team->t.t_nproc;
|
||||
ompt_callbacks.ompt_callback(ompt_callback_implicit_task)(
|
||||
ompt_scope_end, NULL, &(task_info->task_data), ompt_team_size,
|
||||
OMPT_CUR_TASK_INFO(master_th)->thread_num, ompt_task_implicit);
|
||||
}
|
||||
task_info->frame.exit_frame = ompt_data_none;
|
||||
task_info->task_data = ompt_data_none;
|
||||
ompt_parallel_data = *OMPT_CUR_TEAM_DATA(master_th);
|
||||
__ompt_lw_taskteam_unlink(master_th);
|
||||
}
|
||||
#endif
|
||||
// AC: We need to leave the team structure intact at the end of parallel
|
||||
// inside the teams construct, so that at the next parallel same (hot) team
|
||||
// works, only adjust nesting levels
|
||||
|
||||
/* Decrement our nested depth level */
|
||||
team->t.t_level--;
|
||||
team->t.t_active_level--;
|
||||
@@ -2449,8 +2394,8 @@ void __kmp_join_call(ident_t *loc, int gtid
|
||||
|
||||
#if OMPT_SUPPORT
|
||||
if (ompt_enabled.enabled) {
|
||||
__kmp_join_ompt(gtid, master_th, parent_team, &ompt_parallel_data,
|
||||
OMPT_INVOKER(fork_context) | ompt_parallel_team, codeptr);
|
||||
__kmp_join_ompt(gtid, master_th, parent_team, parallel_data, fork_context,
|
||||
codeptr);
|
||||
}
|
||||
#endif
|
||||
|
||||
@@ -2479,14 +2424,12 @@ void __kmp_join_call(ident_t *loc, int gtid
|
||||
if (ompt_enabled.enabled) {
|
||||
ompt_task_info_t *task_info = __ompt_get_task_info_object(0);
|
||||
if (ompt_enabled.ompt_callback_implicit_task) {
|
||||
int flags = (team_microtask == (void *)__kmp_teams_master)
|
||||
? ompt_task_initial
|
||||
: ompt_task_implicit;
|
||||
int ompt_team_size = (flags == ompt_task_initial) ? 0 : team->t.t_nproc;
|
||||
int ompt_team_size = team->t.t_nproc;
|
||||
ompt_callbacks.ompt_callback(ompt_callback_implicit_task)(
|
||||
ompt_scope_end, NULL, &(task_info->task_data), ompt_team_size,
|
||||
OMPT_CUR_TASK_INFO(master_th)->thread_num, flags);
|
||||
OMPT_CUR_TASK_INFO(master_th)->thread_num, ompt_task_implicit); // TODO: Can this be ompt_task_initial?
|
||||
}
|
||||
|
||||
task_info->frame.exit_frame = ompt_data_none;
|
||||
task_info->task_data = ompt_data_none;
|
||||
}
|
||||
@@ -2560,12 +2503,8 @@ void __kmp_join_call(ident_t *loc, int gtid
|
||||
__kmp_release_bootstrap_lock(&__kmp_forkjoin_lock);
|
||||
|
||||
#if OMPT_SUPPORT
|
||||
int flags =
|
||||
OMPT_INVOKER(fork_context) |
|
||||
((team_microtask == (void *)__kmp_teams_master) ? ompt_parallel_league
|
||||
: ompt_parallel_team);
|
||||
if (ompt_enabled.enabled) {
|
||||
__kmp_join_ompt(gtid, master_th, parent_team, parallel_data, flags,
|
||||
__kmp_join_ompt(gtid, master_th, parent_team, parallel_data, fork_context,
|
||||
codeptr);
|
||||
}
|
||||
#endif
|
||||
@@ -4493,7 +4432,7 @@ static void __kmp_initialize_team(kmp_team_t *team, int new_nproc,
|
||||
KF_TRACE(10, ("__kmp_initialize_team: exit: team=%p\n", team));
|
||||
}
|
||||
|
||||
#if (KMP_OS_LINUX || KMP_OS_FREEBSD) && KMP_AFFINITY_SUPPORTED
|
||||
#if KMP_OS_LINUX && KMP_AFFINITY_SUPPORTED
|
||||
/* Sets full mask for thread and returns old mask, no changes to structures. */
|
||||
static void
|
||||
__kmp_set_thread_affinity_mask_full_tmp(kmp_affin_mask_t *old_mask) {
|
||||
@@ -5041,7 +4980,7 @@ __kmp_allocate_team(kmp_root_t *root, int new_nproc, int max_nproc,
|
||||
__kmp_partition_places(team);
|
||||
#endif
|
||||
} else { // team->t.t_nproc < new_nproc
|
||||
#if (KMP_OS_LINUX || KMP_OS_FREEBSD) && KMP_AFFINITY_SUPPORTED
|
||||
#if KMP_OS_LINUX && KMP_AFFINITY_SUPPORTED
|
||||
kmp_affin_mask_t *old_mask;
|
||||
if (KMP_AFFINITY_CAPABLE()) {
|
||||
KMP_CPU_ALLOC(old_mask);
|
||||
@@ -5090,7 +5029,7 @@ __kmp_allocate_team(kmp_root_t *root, int new_nproc, int max_nproc,
|
||||
__kmp_reinitialize_team(team, new_icvs, NULL);
|
||||
}
|
||||
|
||||
#if (KMP_OS_LINUX || KMP_OS_FREEBSD) && KMP_AFFINITY_SUPPORTED
|
||||
#if KMP_OS_LINUX && KMP_AFFINITY_SUPPORTED
|
||||
/* Temporarily set full mask for master thread before creation of
|
||||
workers. The reason is that workers inherit the affinity from master,
|
||||
so if a lot of workers are created on the single core quickly, they
|
||||
@@ -5125,7 +5064,7 @@ __kmp_allocate_team(kmp_root_t *root, int new_nproc, int max_nproc,
|
||||
}
|
||||
}
|
||||
|
||||
#if (KMP_OS_LINUX || KMP_OS_FREEBSD) && KMP_AFFINITY_SUPPORTED
|
||||
#if KMP_OS_LINUX && KMP_AFFINITY_SUPPORTED
|
||||
if (KMP_AFFINITY_CAPABLE()) {
|
||||
/* Restore initial master thread's affinity mask */
|
||||
__kmp_set_system_affinity(old_mask, TRUE);
|
||||
@@ -5661,7 +5600,7 @@ void __kmp_free_thread(kmp_info_t *this_th) {
|
||||
void *__kmp_launch_thread(kmp_info_t *this_thr) {
|
||||
int gtid = this_thr->th.th_info.ds.ds_gtid;
|
||||
/* void *stack_data;*/
|
||||
kmp_team_t **volatile pteam;
|
||||
kmp_team_t *(*volatile pteam);
|
||||
|
||||
KMP_MB();
|
||||
KA_TRACE(10, ("__kmp_launch_thread: T#%d start\n", gtid));
|
||||
@@ -5679,15 +5618,18 @@ void *__kmp_launch_thread(kmp_info_t *this_thr) {
|
||||
this_thr->th.ompt_thread_info.state = ompt_state_overhead;
|
||||
this_thr->th.ompt_thread_info.wait_id = 0;
|
||||
this_thr->th.ompt_thread_info.idle_frame = OMPT_GET_FRAME_ADDRESS(0);
|
||||
this_thr->th.ompt_thread_info.parallel_flags = 0;
|
||||
if (ompt_enabled.ompt_callback_thread_begin) {
|
||||
ompt_callbacks.ompt_callback(ompt_callback_thread_begin)(
|
||||
ompt_thread_worker, thread_data);
|
||||
}
|
||||
this_thr->th.ompt_thread_info.state = ompt_state_idle;
|
||||
}
|
||||
#endif
|
||||
|
||||
#if OMPT_SUPPORT
|
||||
if (ompt_enabled.enabled) {
|
||||
this_thr->th.ompt_thread_info.state = ompt_state_idle;
|
||||
}
|
||||
#endif
|
||||
/* This is the place where threads wait for work */
|
||||
while (!TCR_4(__kmp_global.g.g_done)) {
|
||||
KMP_DEBUG_ASSERT(this_thr == __kmp_threads[gtid]);
|
||||
@@ -5705,7 +5647,7 @@ void *__kmp_launch_thread(kmp_info_t *this_thr) {
|
||||
}
|
||||
#endif
|
||||
|
||||
pteam = &this_thr->th.th_team;
|
||||
pteam = (kmp_team_t * (*))(&this_thr->th.th_team);
|
||||
|
||||
/* have we been allocated? */
|
||||
if (TCR_SYNC_PTR(*pteam) && !TCR_4(__kmp_global.g.g_done)) {
|
||||
@@ -7014,16 +6956,16 @@ int __kmp_invoke_task_func(int gtid) {
|
||||
|
||||
#if OMPT_SUPPORT
|
||||
void *dummy;
|
||||
void **exit_frame_p;
|
||||
void **exit_runtime_p;
|
||||
ompt_data_t *my_task_data;
|
||||
ompt_data_t *my_parallel_data;
|
||||
int ompt_team_size;
|
||||
|
||||
if (ompt_enabled.enabled) {
|
||||
exit_frame_p = &(
|
||||
exit_runtime_p = &(
|
||||
team->t.t_implicit_task_taskdata[tid].ompt_task_info.frame.exit_frame.ptr);
|
||||
} else {
|
||||
exit_frame_p = &dummy;
|
||||
exit_runtime_p = &dummy;
|
||||
}
|
||||
|
||||
my_task_data =
|
||||
@@ -7033,7 +6975,7 @@ int __kmp_invoke_task_func(int gtid) {
|
||||
ompt_team_size = team->t.t_nproc;
|
||||
ompt_callbacks.ompt_callback(ompt_callback_implicit_task)(
|
||||
ompt_scope_begin, my_parallel_data, my_task_data, ompt_team_size,
|
||||
__kmp_tid_from_gtid(gtid), ompt_task_implicit);
|
||||
__kmp_tid_from_gtid(gtid), ompt_task_implicit); // TODO: Can this be ompt_task_initial?
|
||||
OMPT_CUR_TASK_INFO(this_thr)->thread_num = __kmp_tid_from_gtid(gtid);
|
||||
}
|
||||
#endif
|
||||
@@ -7052,12 +6994,11 @@ int __kmp_invoke_task_func(int gtid) {
|
||||
tid, (int)team->t.t_argc, (void **)team->t.t_argv
|
||||
#if OMPT_SUPPORT
|
||||
,
|
||||
exit_frame_p
|
||||
exit_runtime_p
|
||||
#endif
|
||||
);
|
||||
#if OMPT_SUPPORT
|
||||
*exit_frame_p = NULL;
|
||||
this_thr->th.ompt_thread_info.parallel_flags |= ompt_parallel_team;
|
||||
*exit_runtime_p = NULL;
|
||||
#endif
|
||||
|
||||
#if KMP_STATS_ENABLED
|
||||
@@ -7136,22 +7077,7 @@ int __kmp_invoke_teams_master(int gtid) {
|
||||
(void *)__kmp_teams_master);
|
||||
#endif
|
||||
__kmp_run_before_invoked_task(gtid, 0, this_thr, team);
|
||||
#if OMPT_SUPPORT
|
||||
int tid = __kmp_tid_from_gtid(gtid);
|
||||
ompt_data_t *task_data =
|
||||
&team->t.t_implicit_task_taskdata[tid].ompt_task_info.task_data;
|
||||
ompt_data_t *parallel_data = &team->t.ompt_team_info.parallel_data;
|
||||
if (ompt_enabled.ompt_callback_implicit_task) {
|
||||
ompt_callbacks.ompt_callback(ompt_callback_implicit_task)(
|
||||
ompt_scope_begin, parallel_data, task_data, team->t.t_nproc, tid,
|
||||
ompt_task_initial);
|
||||
OMPT_CUR_TASK_INFO(this_thr)->thread_num = tid;
|
||||
}
|
||||
#endif
|
||||
__kmp_teams_master(gtid);
|
||||
#if OMPT_SUPPORT
|
||||
this_thr->th.ompt_thread_info.parallel_flags |= ompt_parallel_league;
|
||||
#endif
|
||||
__kmp_run_after_invoked_task(gtid, 0, this_thr, team);
|
||||
return 1;
|
||||
}
|
||||
@@ -7192,32 +7118,19 @@ void __kmp_push_num_teams(ident_t *id, int gtid, int num_teams,
|
||||
thr->th.th_set_nproc = thr->th.th_teams_size.nteams = num_teams;
|
||||
|
||||
// Remember the number of threads for inner parallel regions
|
||||
if (!TCR_4(__kmp_init_middle))
|
||||
__kmp_middle_initialize(); // get internal globals calculated
|
||||
KMP_DEBUG_ASSERT(__kmp_avail_proc);
|
||||
KMP_DEBUG_ASSERT(__kmp_dflt_team_nth);
|
||||
if (num_threads == 0) {
|
||||
if (!TCR_4(__kmp_init_middle))
|
||||
__kmp_middle_initialize(); // get __kmp_avail_proc calculated
|
||||
num_threads = __kmp_avail_proc / num_teams;
|
||||
// adjust num_threads w/o warning as it is not user setting
|
||||
// num_threads = min(num_threads, nthreads-var, thread-limit-var)
|
||||
// no thread_limit clause specified - do not change thread-limit-var ICV
|
||||
if (num_threads > __kmp_dflt_team_nth) {
|
||||
num_threads = __kmp_dflt_team_nth; // honor nthreads-var ICV
|
||||
}
|
||||
if (num_threads > thr->th.th_current_task->td_icvs.thread_limit) {
|
||||
num_threads = thr->th.th_current_task->td_icvs.thread_limit;
|
||||
} // prevent team size to exceed thread-limit-var
|
||||
if (num_teams * num_threads > __kmp_teams_max_nth) {
|
||||
// adjust num_threads w/o warning as it is not user setting
|
||||
num_threads = __kmp_teams_max_nth / num_teams;
|
||||
}
|
||||
} else {
|
||||
// This thread will be the master of the league masters
|
||||
// Store new thread limit; old limit is saved in th_cg_roots list
|
||||
thr->th.th_current_task->td_icvs.thread_limit = num_threads;
|
||||
// num_threads = min(num_threads, nthreads-var)
|
||||
if (num_threads > __kmp_dflt_team_nth) {
|
||||
num_threads = __kmp_dflt_team_nth; // honor nthreads-var ICV
|
||||
}
|
||||
|
||||
if (num_teams * num_threads > __kmp_teams_max_nth) {
|
||||
int new_threads = __kmp_teams_max_nth / num_teams;
|
||||
if (!__kmp_reserve_warn) { // user asked for too many threads
|
||||
@@ -8110,8 +8023,7 @@ __kmp_determine_reduction_method(
|
||||
|
||||
int atomic_available = FAST_REDUCTION_ATOMIC_METHOD_GENERATED;
|
||||
|
||||
#if KMP_ARCH_X86_64 || KMP_ARCH_PPC64 || KMP_ARCH_AARCH64 || \
|
||||
KMP_ARCH_MIPS64 || KMP_ARCH_RISCV64
|
||||
#if KMP_ARCH_X86_64 || KMP_ARCH_PPC64 || KMP_ARCH_AARCH64 || KMP_ARCH_MIPS64
|
||||
|
||||
#if KMP_OS_LINUX || KMP_OS_DRAGONFLY || KMP_OS_FREEBSD || KMP_OS_NETBSD || \
|
||||
KMP_OS_OPENBSD || KMP_OS_WINDOWS || KMP_OS_DARWIN || KMP_OS_HURD
|
||||
|
||||
@@ -164,7 +164,7 @@ void *kmp_aligned_malloc(size_t sz, size_t a) {
|
||||
#if KMP_OS_WINDOWS
|
||||
res = _aligned_malloc(sz, a);
|
||||
#else
|
||||
if ((err = posix_memalign(&res, a, sz))) {
|
||||
if (err = posix_memalign(&res, a, sz)) {
|
||||
errno = err; // can be EINVAL or ENOMEM
|
||||
res = NULL;
|
||||
}
|
||||
@@ -277,7 +277,7 @@ void __kmps_get_schedule(kmp_sched_t *kind, int *modifier) {
|
||||
|
||||
kmp_proc_bind_t __kmps_get_proc_bind(void) {
|
||||
i;
|
||||
return proc_bind_false;
|
||||
return 0;
|
||||
} // __kmps_get_proc_bind
|
||||
|
||||
double __kmps_get_wtime(void) {
|
||||
|
||||
@@ -54,64 +54,12 @@ static inline kmp_depnode_t *__kmp_node_ref(kmp_depnode_t *node) {
|
||||
|
||||
enum { KMP_DEPHASH_OTHER_SIZE = 97, KMP_DEPHASH_MASTER_SIZE = 997 };
|
||||
|
||||
size_t sizes[] = { 997, 2003, 4001, 8191, 16001, 32003, 64007, 131071, 270029 };
|
||||
const size_t MAX_GEN = 8;
|
||||
|
||||
static inline kmp_int32 __kmp_dephash_hash(kmp_intptr_t addr, size_t hsize) {
|
||||
// TODO alternate to try: set = (((Addr64)(addrUsefulBits * 9.618)) %
|
||||
// m_num_sets );
|
||||
return ((addr >> 6) ^ (addr >> 2)) % hsize;
|
||||
}
|
||||
|
||||
static kmp_dephash_t *__kmp_dephash_extend(kmp_info_t *thread,
|
||||
kmp_dephash_t *current_dephash) {
|
||||
kmp_dephash_t *h;
|
||||
|
||||
size_t gen = current_dephash->generation + 1;
|
||||
if (gen >= MAX_GEN)
|
||||
return current_dephash;
|
||||
size_t new_size = sizes[gen];
|
||||
|
||||
kmp_int32 size_to_allocate =
|
||||
new_size * sizeof(kmp_dephash_entry_t *) + sizeof(kmp_dephash_t);
|
||||
|
||||
#if USE_FAST_MEMORY
|
||||
h = (kmp_dephash_t *)__kmp_fast_allocate(thread, size_to_allocate);
|
||||
#else
|
||||
h = (kmp_dephash_t *)__kmp_thread_malloc(thread, size_to_allocate);
|
||||
#endif
|
||||
|
||||
h->size = new_size;
|
||||
h->nelements = current_dephash->nelements;
|
||||
h->buckets = (kmp_dephash_entry **)(h + 1);
|
||||
h->generation = gen;
|
||||
|
||||
// insert existing elements in the new table
|
||||
for (size_t i = 0; i < current_dephash->size; i++) {
|
||||
kmp_dephash_entry_t *next;
|
||||
for (kmp_dephash_entry_t *entry = current_dephash->buckets[i]; entry; entry = next) {
|
||||
next = entry->next_in_bucket;
|
||||
// Compute the new hash using the new size, and insert the entry in
|
||||
// the new bucket.
|
||||
kmp_int32 new_bucket = __kmp_dephash_hash(entry->addr, h->size);
|
||||
if (entry->next_in_bucket) {
|
||||
h->nconflicts++;
|
||||
}
|
||||
entry->next_in_bucket = h->buckets[new_bucket];
|
||||
h->buckets[new_bucket] = entry;
|
||||
}
|
||||
}
|
||||
|
||||
// Free old hash table
|
||||
#if USE_FAST_MEMORY
|
||||
__kmp_fast_free(thread, current_dephash);
|
||||
#else
|
||||
__kmp_thread_free(thread, current_dephash);
|
||||
#endif
|
||||
|
||||
return h;
|
||||
}
|
||||
|
||||
static kmp_dephash_t *__kmp_dephash_create(kmp_info_t *thread,
|
||||
kmp_taskdata_t *current_task) {
|
||||
kmp_dephash_t *h;
|
||||
@@ -133,9 +81,10 @@ static kmp_dephash_t *__kmp_dephash_create(kmp_info_t *thread,
|
||||
#endif
|
||||
h->size = h_size;
|
||||
|
||||
h->generation = 0;
|
||||
#ifdef KMP_DEBUG
|
||||
h->nelements = 0;
|
||||
h->nconflicts = 0;
|
||||
#endif
|
||||
h->buckets = (kmp_dephash_entry **)(h + 1);
|
||||
|
||||
for (size_t i = 0; i < h_size; i++)
|
||||
@@ -148,13 +97,7 @@ static kmp_dephash_t *__kmp_dephash_create(kmp_info_t *thread,
|
||||
#define ENTRY_LAST_MTXS 1
|
||||
|
||||
static kmp_dephash_entry *
|
||||
__kmp_dephash_find(kmp_info_t *thread, kmp_dephash_t **hash, kmp_intptr_t addr) {
|
||||
kmp_dephash_t *h = *hash;
|
||||
if (h->nelements != 0
|
||||
&& h->nconflicts/h->size >= 1) {
|
||||
*hash = __kmp_dephash_extend(thread, h);
|
||||
h = *hash;
|
||||
}
|
||||
__kmp_dephash_find(kmp_info_t *thread, kmp_dephash_t *h, kmp_intptr_t addr) {
|
||||
kmp_int32 bucket = __kmp_dephash_hash(addr, h->size);
|
||||
|
||||
kmp_dephash_entry_t *entry;
|
||||
@@ -179,9 +122,11 @@ __kmp_dephash_find(kmp_info_t *thread, kmp_dephash_t **hash, kmp_intptr_t addr)
|
||||
entry->mtx_lock = NULL;
|
||||
entry->next_in_bucket = h->buckets[bucket];
|
||||
h->buckets[bucket] = entry;
|
||||
#ifdef KMP_DEBUG
|
||||
h->nelements++;
|
||||
if (entry->next_in_bucket)
|
||||
h->nconflicts++;
|
||||
#endif
|
||||
}
|
||||
return entry;
|
||||
}
|
||||
@@ -287,7 +232,7 @@ static inline kmp_int32 __kmp_depnode_link_successor(kmp_int32 gtid,
|
||||
|
||||
template <bool filter>
|
||||
static inline kmp_int32
|
||||
__kmp_process_deps(kmp_int32 gtid, kmp_depnode_t *node, kmp_dephash_t **hash,
|
||||
__kmp_process_deps(kmp_int32 gtid, kmp_depnode_t *node, kmp_dephash_t *hash,
|
||||
bool dep_barrier, kmp_int32 ndeps,
|
||||
kmp_depend_info_t *dep_list, kmp_task_t *task) {
|
||||
KA_TRACE(30, ("__kmp_process_deps<%d>: T#%d processing %d dependencies : "
|
||||
@@ -407,7 +352,7 @@ __kmp_process_deps(kmp_int32 gtid, kmp_depnode_t *node, kmp_dephash_t **hash,
|
||||
|
||||
// returns true if the task has any outstanding dependence
|
||||
static bool __kmp_check_deps(kmp_int32 gtid, kmp_depnode_t *node,
|
||||
kmp_task_t *task, kmp_dephash_t **hash,
|
||||
kmp_task_t *task, kmp_dephash_t *hash,
|
||||
bool dep_barrier, kmp_int32 ndeps,
|
||||
kmp_depend_info_t *dep_list,
|
||||
kmp_int32 ndeps_noalias,
|
||||
@@ -607,7 +552,7 @@ kmp_int32 __kmpc_omp_task_with_deps(ident_t *loc_ref, kmp_int32 gtid,
|
||||
__kmp_init_node(node);
|
||||
new_taskdata->td_depnode = node;
|
||||
|
||||
if (__kmp_check_deps(gtid, node, new_task, ¤t_task->td_dephash,
|
||||
if (__kmp_check_deps(gtid, node, new_task, current_task->td_dephash,
|
||||
NO_DEP_BARRIER, ndeps, dep_list, ndeps_noalias,
|
||||
noalias_dep_list)) {
|
||||
KA_TRACE(10, ("__kmpc_omp_task_with_deps(exit): T#%d task had blocking "
|
||||
@@ -688,7 +633,7 @@ void __kmpc_omp_wait_deps(ident_t *loc_ref, kmp_int32 gtid, kmp_int32 ndeps,
|
||||
kmp_depnode_t node = {0};
|
||||
__kmp_init_node(&node);
|
||||
|
||||
if (!__kmp_check_deps(gtid, &node, NULL, ¤t_task->td_dephash,
|
||||
if (!__kmp_check_deps(gtid, &node, NULL, current_task->td_dephash,
|
||||
DEP_BARRIER, ndeps, dep_list, ndeps_noalias,
|
||||
noalias_dep_list)) {
|
||||
KA_TRACE(10, ("__kmpc_omp_wait_deps(exit): T#%d has no blocking "
|
||||
|
||||
@@ -140,11 +140,8 @@ static void __ompt_implicit_task_end(kmp_info_t *this_thr,
|
||||
#endif
|
||||
if (!KMP_MASTER_TID(ds_tid)) {
|
||||
if (ompt_enabled.ompt_callback_implicit_task) {
|
||||
int flags = this_thr->th.ompt_thread_info.parallel_flags;
|
||||
flags = (flags & ompt_parallel_league) ? ompt_task_initial
|
||||
: ompt_task_implicit;
|
||||
ompt_callbacks.ompt_callback(ompt_callback_implicit_task)(
|
||||
ompt_scope_end, NULL, tId, 0, ds_tid, flags);
|
||||
ompt_scope_end, NULL, tId, 0, ds_tid, ompt_task_implicit);
|
||||
}
|
||||
// return to idle state
|
||||
this_thr->th.ompt_thread_info.state = ompt_state_idle;
|
||||
|
||||
@@ -430,8 +430,10 @@ OMPT_API_ROUTINE ompt_set_result_t ompt_set_callback(ompt_callbacks_t which,
|
||||
|
||||
#define ompt_event_macro(event_name, callback_type, event_id) \
|
||||
case event_name: \
|
||||
ompt_callbacks.ompt_callback(event_name) = (callback_type)callback; \
|
||||
ompt_enabled.event_name = (callback != 0); \
|
||||
if (ompt_event_implementation_status(event_name)) { \
|
||||
ompt_callbacks.ompt_callback(event_name) = (callback_type)callback; \
|
||||
ompt_enabled.event_name = (callback != 0); \
|
||||
} \
|
||||
if (callback) \
|
||||
return ompt_event_implementation_status(event_name); \
|
||||
else \
|
||||
@@ -454,15 +456,16 @@ OMPT_API_ROUTINE int ompt_get_callback(ompt_callbacks_t which,
|
||||
switch (which) {
|
||||
|
||||
#define ompt_event_macro(event_name, callback_type, event_id) \
|
||||
case event_name: { \
|
||||
ompt_callback_t mycb = \
|
||||
(ompt_callback_t)ompt_callbacks.ompt_callback(event_name); \
|
||||
if (ompt_enabled.event_name && mycb) { \
|
||||
*callback = mycb; \
|
||||
return ompt_get_callback_success; \
|
||||
case event_name: \
|
||||
if (ompt_event_implementation_status(event_name)) { \
|
||||
ompt_callback_t mycb = \
|
||||
(ompt_callback_t)ompt_callbacks.ompt_callback(event_name); \
|
||||
if (ompt_enabled.event_name && mycb) { \
|
||||
*callback = mycb; \
|
||||
return ompt_get_callback_success; \
|
||||
} \
|
||||
} \
|
||||
return ompt_get_callback_failure; \
|
||||
}
|
||||
return ompt_get_callback_failure;
|
||||
|
||||
FOREACH_OMPT_EVENT(ompt_event_macro)
|
||||
|
||||
|
||||
@@ -81,7 +81,6 @@ typedef struct {
|
||||
ompt_state_t state;
|
||||
ompt_wait_id_t wait_id;
|
||||
int ompt_task_yielded;
|
||||
int parallel_flags; // information for the last parallel region invoked
|
||||
void *idle_frame;
|
||||
} ompt_thread_info_t;
|
||||
|
||||
|
||||
@@ -269,11 +269,10 @@ void __ompt_lw_taskteam_init(ompt_lw_taskteam_t *lwt, kmp_info_t *thr, int gtid,
|
||||
}
|
||||
|
||||
void __ompt_lw_taskteam_link(ompt_lw_taskteam_t *lwt, kmp_info_t *thr,
|
||||
int on_heap, bool always) {
|
||||
int on_heap) {
|
||||
ompt_lw_taskteam_t *link_lwt = lwt;
|
||||
if (always ||
|
||||
thr->th.th_team->t.t_serialized >
|
||||
1) { // we already have a team, so link the new team and swap values
|
||||
if (thr->th.th_team->t.t_serialized >
|
||||
1) { // we already have a team, so link the new team and swap values
|
||||
if (on_heap) { // the lw_taskteam cannot stay on stack, allocate it on heap
|
||||
link_lwt =
|
||||
(ompt_lw_taskteam_t *)__kmp_allocate(sizeof(ompt_lw_taskteam_t));
|
||||
|
||||
@@ -26,7 +26,7 @@ void __ompt_lw_taskteam_init(ompt_lw_taskteam_t *lwt, kmp_info_t *thr,
|
||||
int gtid, ompt_data_t *ompt_pid, void *codeptr);
|
||||
|
||||
void __ompt_lw_taskteam_link(ompt_lw_taskteam_t *lwt, kmp_info_t *thr,
|
||||
int on_heap, bool always = false);
|
||||
int on_heap);
|
||||
|
||||
void __ompt_lw_taskteam_unlink(kmp_info_t *thr);
|
||||
|
||||
|
||||
+1
-9
@@ -161,10 +161,6 @@
|
||||
# define ITT_ARCH_MIPS64 6
|
||||
#endif /* ITT_ARCH_MIPS64 */
|
||||
|
||||
#ifndef ITT_ARCH_RISCV64
|
||||
# define ITT_ARCH_RISCV64 7
|
||||
#endif /* ITT_ARCH_RISCV64 */
|
||||
|
||||
#ifndef ITT_ARCH
|
||||
# if defined _M_IX86 || defined __i386__
|
||||
# define ITT_ARCH ITT_ARCH_IA32
|
||||
@@ -182,8 +178,6 @@
|
||||
# define ITT_ARCH ITT_ARCH_MIPS
|
||||
# elif defined __mips__ && defined __mips64
|
||||
# define ITT_ARCH ITT_ARCH_MIPS64
|
||||
# elif defined __riscv && __riscv_xlen == 64
|
||||
# define ITT_ARCH ITT_ARCH_RISCV64
|
||||
# endif
|
||||
#endif
|
||||
|
||||
@@ -336,9 +330,7 @@ ITT_INLINE long __TBB_machine_fetchadd4(volatile void* ptr, long addend)
|
||||
: "memory");
|
||||
return result;
|
||||
}
|
||||
#elif ITT_ARCH == ITT_ARCH_ARM || ITT_ARCH == ITT_ARCH_PPC64 || \
|
||||
ITT_ARCH == ITT_ARCH_AARCH64 || ITT_ARCH == ITT_ARCH_MIPS || \
|
||||
ITT_ARCH == ITT_ARCH_MIPS64 || ITT_ARCH == ITT_ARCH_RISCV64
|
||||
#elif ITT_ARCH==ITT_ARCH_ARM || ITT_ARCH==ITT_ARCH_PPC64 || ITT_ARCH==ITT_ARCH_AARCH64 || ITT_ARCH==ITT_ARCH_MIPS || ITT_ARCH==ITT_ARCH_MIPS64
|
||||
#define __TBB_machine_fetchadd4(addr, val) __sync_fetch_and_add(addr, val)
|
||||
#endif /* ITT_ARCH==ITT_ARCH_IA64 */
|
||||
#ifndef ITT_SIMPLE_INIT
|
||||
|
||||
+3
-3
@@ -8,7 +8,6 @@
|
||||
//===----------------------------------------------------------------------===//
|
||||
|
||||
#include "kmp_config.h"
|
||||
#include "kmp_os.h"
|
||||
#include "ittnotify_config.h"
|
||||
|
||||
#if ITT_PLATFORM==ITT_PLATFORM_WIN
|
||||
@@ -227,6 +226,8 @@ static __itt_api_info api_list[] = {
|
||||
#pragma warning(pop)
|
||||
#endif /* ITT_PLATFORM==ITT_PLATFORM_WIN */
|
||||
|
||||
static const char dll_path[PATH_MAX] = { 0 };
|
||||
|
||||
/* static part descriptor which handles. all notification api attributes. */
|
||||
__itt_global _N_(_ittapi_global) = {
|
||||
ITT_MAGIC, /* identification info */
|
||||
@@ -237,7 +238,7 @@ __itt_global _N_(_ittapi_global) = {
|
||||
MUTEX_INITIALIZER, /* mutex */
|
||||
NULL, /* dynamic library handle */
|
||||
NULL, /* error_handler */
|
||||
NULL, /* dll_path_ptr */
|
||||
(const char**)&dll_path, /* dll_path_ptr */
|
||||
(__itt_api_info*)&api_list, /* api_list_ptr */
|
||||
NULL, /* next __itt_global */
|
||||
NULL, /* thread_list */
|
||||
@@ -1097,7 +1098,6 @@ ITT_EXTERN_C int _N_(init_ittlib)(const char* lib_name, __itt_group_id init_grou
|
||||
switch (lib_version) {
|
||||
case 0:
|
||||
groups = __itt_group_legacy;
|
||||
KMP_FALLTHROUGH();
|
||||
case 1:
|
||||
/* Fill all pointers from dynamic library */
|
||||
for (i = 0; _N_(_ittapi_global).api_list_ptr[i].name != NULL; i++)
|
||||
+28
-235
@@ -495,21 +495,13 @@ __kmp_unnamed_critical_addr:
|
||||
# endif /* !KMP_ASM_INTRINS */
|
||||
|
||||
//------------------------------------------------------------------------
|
||||
// int
|
||||
// __kmp_invoke_microtask( void (*pkfn) (int gtid, int tid, ...),
|
||||
// int gtid, int tid,
|
||||
// int argc, void *p_argv[]
|
||||
// #if OMPT_SUPPORT
|
||||
// ,
|
||||
// void **exit_frame_ptr
|
||||
// #endif
|
||||
// ) {
|
||||
// #if OMPT_SUPPORT
|
||||
// *exit_frame_ptr = OMPT_GET_FRAME_ADDRESS(0);
|
||||
// #endif
|
||||
// typedef void (*microtask_t)( int *gtid, int *tid, ... );
|
||||
//
|
||||
// (*pkfn)( & gtid, & tid, argv[0], ... );
|
||||
// return 1;
|
||||
// int
|
||||
// __kmp_invoke_microtask( microtask_t pkfn, int gtid, int tid,
|
||||
// int argc, void *p_argv[] ) {
|
||||
// (*pkfn)( & gtid, & gtid, argv[0], ... );
|
||||
// return 1;
|
||||
// }
|
||||
|
||||
// -- Begin __kmp_invoke_microtask
|
||||
@@ -999,21 +991,14 @@ KMP_LABEL(invoke_3):
|
||||
# endif /* !KMP_ASM_INTRINS */
|
||||
|
||||
//------------------------------------------------------------------------
|
||||
// typedef void (*microtask_t)( int *gtid, int *tid, ... );
|
||||
//
|
||||
// int
|
||||
// __kmp_invoke_microtask( void (*pkfn) (int gtid, int tid, ...),
|
||||
// int gtid, int tid,
|
||||
// int argc, void *p_argv[]
|
||||
// #if OMPT_SUPPORT
|
||||
// ,
|
||||
// void **exit_frame_ptr
|
||||
// #endif
|
||||
// ) {
|
||||
// #if OMPT_SUPPORT
|
||||
// *exit_frame_ptr = OMPT_GET_FRAME_ADDRESS(0);
|
||||
// #endif
|
||||
//
|
||||
// (*pkfn)( & gtid, & tid, argv[0], ... );
|
||||
// return 1;
|
||||
// int gtid, int tid,
|
||||
// int argc, void *p_argv[] ) {
|
||||
// (*pkfn)( & gtid, & tid, argv[0], ... );
|
||||
// return 1;
|
||||
// }
|
||||
//
|
||||
// note: at call to pkfn must have %rsp 128-byte aligned for compiler
|
||||
@@ -1207,27 +1192,15 @@ KMP_LABEL(kmp_1_exit):
|
||||
#if (KMP_OS_LINUX || KMP_OS_DARWIN) && KMP_ARCH_AARCH64
|
||||
|
||||
//------------------------------------------------------------------------
|
||||
//
|
||||
// typedef void (*microtask_t)( int *gtid, int *tid, ... );
|
||||
//
|
||||
// int
|
||||
// __kmp_invoke_microtask( void (*pkfn) (int gtid, int tid, ...),
|
||||
// int gtid, int tid,
|
||||
// int argc, void *p_argv[]
|
||||
// #if OMPT_SUPPORT
|
||||
// ,
|
||||
// void **exit_frame_ptr
|
||||
// #endif
|
||||
// ) {
|
||||
// #if OMPT_SUPPORT
|
||||
// *exit_frame_ptr = OMPT_GET_FRAME_ADDRESS(0);
|
||||
// #endif
|
||||
//
|
||||
// (*pkfn)( & gtid, & tid, argv[0], ... );
|
||||
//
|
||||
// // FIXME: This is done at call-site and can be removed here.
|
||||
// #if OMPT_SUPPORT
|
||||
// *exit_frame_ptr = 0;
|
||||
// #endif
|
||||
//
|
||||
// return 1;
|
||||
// int gtid, int tid,
|
||||
// int argc, void *p_argv[] ) {
|
||||
// (*pkfn)( & gtid, & tid, argv[0], ... );
|
||||
// return 1;
|
||||
// }
|
||||
//
|
||||
// parameters:
|
||||
@@ -1333,27 +1306,15 @@ KMP_LABEL(kmp_1):
|
||||
#if KMP_ARCH_PPC64
|
||||
|
||||
//------------------------------------------------------------------------
|
||||
//
|
||||
// typedef void (*microtask_t)( int *gtid, int *tid, ... );
|
||||
//
|
||||
// int
|
||||
// __kmp_invoke_microtask( void (*pkfn) (int gtid, int tid, ...),
|
||||
// int gtid, int tid,
|
||||
// int argc, void *p_argv[]
|
||||
// #if OMPT_SUPPORT
|
||||
// ,
|
||||
// void **exit_frame_ptr
|
||||
// #endif
|
||||
// ) {
|
||||
// #if OMPT_SUPPORT
|
||||
// *exit_frame_ptr = OMPT_GET_FRAME_ADDRESS(0);
|
||||
// #endif
|
||||
//
|
||||
// (*pkfn)( & gtid, & tid, argv[0], ... );
|
||||
//
|
||||
// // FIXME: This is done at call-site and can be removed here.
|
||||
// #if OMPT_SUPPORT
|
||||
// *exit_frame_ptr = 0;
|
||||
// #endif
|
||||
//
|
||||
// return 1;
|
||||
// int gtid, int tid,
|
||||
// int argc, void *p_argv[] ) {
|
||||
// (*pkfn)( & gtid, & tid, argv[0], ... );
|
||||
// return 1;
|
||||
// }
|
||||
//
|
||||
// parameters:
|
||||
@@ -1563,173 +1524,6 @@ __kmp_invoke_microtask:
|
||||
|
||||
#endif /* KMP_ARCH_PPC64 */
|
||||
|
||||
#if KMP_ARCH_RISCV64
|
||||
|
||||
//------------------------------------------------------------------------
|
||||
//
|
||||
// typedef void (*microtask_t)(int *gtid, int *tid, ...);
|
||||
//
|
||||
// int __kmp_invoke_microtask(microtask_t pkfn, int gtid, int tid, int argc,
|
||||
// void *p_argv[]
|
||||
// #if OMPT_SUPPORT
|
||||
// ,
|
||||
// void **exit_frame_ptr
|
||||
// #endif
|
||||
// ) {
|
||||
// #if OMPT_SUPPORT
|
||||
// *exit_frame_ptr = OMPT_GET_FRAME_ADDRESS(0);
|
||||
// #endif
|
||||
//
|
||||
// (*pkfn)(>id, &tid, argv[0], ...);
|
||||
//
|
||||
// return 1;
|
||||
// }
|
||||
//
|
||||
// Parameters:
|
||||
// a0: pkfn
|
||||
// a1: gtid
|
||||
// a2: tid
|
||||
// a3: argc
|
||||
// a4: p_argv
|
||||
// a5: exit_frame_ptr
|
||||
//
|
||||
// Locals:
|
||||
// __gtid: gtid param pushed on stack so can pass >id to pkfn
|
||||
// __tid: tid param pushed on stack so can pass &tid to pkfn
|
||||
//
|
||||
// Temp. registers:
|
||||
//
|
||||
// t0: used to calculate the dynamic stack size / used to hold pkfn address
|
||||
// t1: used as temporary for stack placement calculation
|
||||
// t2: used as temporary for stack arguments
|
||||
// t3: used as temporary for number of remaining pkfn parms
|
||||
// t4: used to traverse p_argv array
|
||||
//
|
||||
// return: a0 (always 1/TRUE)
|
||||
//
|
||||
|
||||
__gtid = -20
|
||||
__tid = -24
|
||||
|
||||
// -- Begin __kmp_invoke_microtask
|
||||
// mark_begin;
|
||||
.text
|
||||
.globl __kmp_invoke_microtask
|
||||
.p2align 1
|
||||
.type __kmp_invoke_microtask,@function
|
||||
__kmp_invoke_microtask:
|
||||
.cfi_startproc
|
||||
|
||||
// First, save ra and fp
|
||||
addi sp, sp, -16
|
||||
sd ra, 8(sp)
|
||||
sd fp, 0(sp)
|
||||
addi fp, sp, 16
|
||||
.cfi_def_cfa fp, 0
|
||||
.cfi_offset ra, -8
|
||||
.cfi_offset fp, -16
|
||||
|
||||
// Compute the dynamic stack size:
|
||||
//
|
||||
// - We need 8 bytes for storing 'gtid' and 'tid', so we can pass them by
|
||||
// reference
|
||||
// - We need 8 bytes for each argument that cannot be passed to the 'pkfn'
|
||||
// function by register. Given that we have 8 of such registers (a[0-7])
|
||||
// and two + 'argc' arguments (consider >id and &tid), we need to
|
||||
// reserve max(0, argc - 6)*8 extra bytes
|
||||
//
|
||||
// The total number of bytes is then max(0, argc - 6)*8 + 8
|
||||
|
||||
// Compute max(0, argc - 6) using the following bithack:
|
||||
// max(0, x) = x - (x & (x >> 31)), where x := argc - 6
|
||||
// Source: http://graphics.stanford.edu/~seander/bithacks.html#IntegerMinOrMax
|
||||
addi t0, a3, -6
|
||||
srai t1, t0, 31
|
||||
and t1, t0, t1
|
||||
sub t0, t0, t1
|
||||
|
||||
addi t0, t0, 1
|
||||
|
||||
slli t0, t0, 3
|
||||
sub sp, sp, t0
|
||||
|
||||
// Align the stack to 16 bytes
|
||||
andi sp, sp, -16
|
||||
|
||||
mv t0, a0
|
||||
mv t3, a3
|
||||
mv t4, a4
|
||||
|
||||
#if OMPT_SUPPORT
|
||||
// Save frame pointer into exit_frame
|
||||
sd fp, 0(a5)
|
||||
#endif
|
||||
|
||||
// Prepare arguments for the pkfn function (first 8 using a0-a7 registers)
|
||||
|
||||
sw a1, __gtid(fp)
|
||||
sw a2, __tid(fp)
|
||||
|
||||
addi a0, fp, __gtid
|
||||
addi a1, fp, __tid
|
||||
|
||||
beqz t3, .L_kmp_3
|
||||
ld a2, 0(t4)
|
||||
|
||||
addi t3, t3, -1
|
||||
beqz t3, .L_kmp_3
|
||||
ld a3, 8(t4)
|
||||
|
||||
addi t3, t3, -1
|
||||
beqz t3, .L_kmp_3
|
||||
ld a4, 16(t4)
|
||||
|
||||
addi t3, t3, -1
|
||||
beqz t3, .L_kmp_3
|
||||
ld a5, 24(t4)
|
||||
|
||||
addi t3, t3, -1
|
||||
beqz t3, .L_kmp_3
|
||||
ld a6, 32(t4)
|
||||
|
||||
addi t3, t3, -1
|
||||
beqz t3, .L_kmp_3
|
||||
ld a7, 40(t4)
|
||||
|
||||
// Prepare any additional argument passed through the stack
|
||||
addi t4, t4, 48
|
||||
mv t1, sp
|
||||
j .L_kmp_2
|
||||
.L_kmp_1:
|
||||
ld t2, 0(t4)
|
||||
sd t2, 0(t1)
|
||||
addi t4, t4, 8
|
||||
addi t1, t1, 8
|
||||
.L_kmp_2:
|
||||
addi t3, t3, -1
|
||||
bnez t3, .L_kmp_1
|
||||
|
||||
.L_kmp_3:
|
||||
// Call pkfn function
|
||||
jalr t0
|
||||
|
||||
// Restore stack and return
|
||||
|
||||
addi a0, zero, 1
|
||||
|
||||
addi sp, fp, -16
|
||||
ld fp, 0(sp)
|
||||
ld ra, 8(sp)
|
||||
addi sp, sp, 16
|
||||
ret
|
||||
.Lfunc_end0:
|
||||
.size __kmp_invoke_microtask, .Lfunc_end0-__kmp_invoke_microtask
|
||||
.cfi_endproc
|
||||
|
||||
// -- End __kmp_invoke_microtask
|
||||
|
||||
#endif /* KMP_ARCH_RISCV64 */
|
||||
|
||||
#if KMP_ARCH_ARM || KMP_ARCH_MIPS
|
||||
.data
|
||||
.comm .gomp_critical_user_,32,8
|
||||
@@ -1741,7 +1535,7 @@ __kmp_unnamed_critical_addr:
|
||||
.size __kmp_unnamed_critical_addr,4
|
||||
#endif /* KMP_ARCH_ARM */
|
||||
|
||||
#if KMP_ARCH_PPC64 || KMP_ARCH_AARCH64 || KMP_ARCH_MIPS64 || KMP_ARCH_RISCV64
|
||||
#if KMP_ARCH_PPC64 || KMP_ARCH_AARCH64 || KMP_ARCH_MIPS64
|
||||
.data
|
||||
.comm .gomp_critical_user_,32,8
|
||||
.data
|
||||
@@ -1750,8 +1544,7 @@ __kmp_unnamed_critical_addr:
|
||||
__kmp_unnamed_critical_addr:
|
||||
.8byte .gomp_critical_user_
|
||||
.size __kmp_unnamed_critical_addr,8
|
||||
#endif /* KMP_ARCH_PPC64 || KMP_ARCH_AARCH64 || KMP_ARCH_MIPS64 ||
|
||||
KMP_ARCH_RISCV64 */
|
||||
#endif /* KMP_ARCH_PPC64 || KMP_ARCH_AARCH64 */
|
||||
|
||||
#if KMP_OS_LINUX
|
||||
# if KMP_ARCH_ARM
|
||||
|
||||
@@ -50,9 +50,6 @@
|
||||
#include <mach/mach.h>
|
||||
#include <sys/sysctl.h>
|
||||
#elif KMP_OS_DRAGONFLY || KMP_OS_FREEBSD
|
||||
#include <sys/types.h>
|
||||
#include <sys/sysctl.h>
|
||||
#include <sys/user.h>
|
||||
#include <pthread_np.h>
|
||||
#elif KMP_OS_NETBSD
|
||||
#include <sys/types.h>
|
||||
@@ -100,7 +97,7 @@ static void __kmp_print_cond(char *buffer, kmp_cond_align_t *cond) {
|
||||
}
|
||||
#endif
|
||||
|
||||
#if ((KMP_OS_LINUX || KMP_OS_FREEBSD) && KMP_AFFINITY_SUPPORTED)
|
||||
#if (KMP_OS_LINUX && KMP_AFFINITY_SUPPORTED)
|
||||
|
||||
/* Affinity support */
|
||||
|
||||
@@ -122,21 +119,16 @@ void __kmp_affinity_bind_thread(int which) {
|
||||
void __kmp_affinity_determine_capable(const char *env_var) {
|
||||
// Check and see if the OS supports thread affinity.
|
||||
|
||||
#if KMP_OS_LINUX
|
||||
#define KMP_CPU_SET_SIZE_LIMIT (1024 * 1024)
|
||||
#elif KMP_OS_FREEBSD
|
||||
#define KMP_CPU_SET_SIZE_LIMIT (sizeof(cpuset_t))
|
||||
#endif
|
||||
|
||||
|
||||
#if KMP_OS_LINUX
|
||||
// If Linux* OS:
|
||||
// If the syscall fails or returns a suggestion for the size,
|
||||
// then we don't have to search for an appropriate size.
|
||||
int gCode;
|
||||
int sCode;
|
||||
unsigned char *buf;
|
||||
buf = (unsigned char *)KMP_INTERNAL_MALLOC(KMP_CPU_SET_SIZE_LIMIT);
|
||||
|
||||
// If Linux* OS:
|
||||
// If the syscall fails or returns a suggestion for the size,
|
||||
// then we don't have to search for an appropriate size.
|
||||
gCode = syscall(__NR_sched_getaffinity, 0, KMP_CPU_SET_SIZE_LIMIT, buf);
|
||||
KA_TRACE(30, ("__kmp_affinity_determine_capable: "
|
||||
"initial getaffinity call returned %d errno = %d\n",
|
||||
@@ -275,23 +267,6 @@ void __kmp_affinity_determine_capable(const char *env_var) {
|
||||
}
|
||||
}
|
||||
}
|
||||
#elif KMP_OS_FREEBSD
|
||||
int gCode;
|
||||
unsigned char *buf;
|
||||
buf = (unsigned char *)KMP_INTERNAL_MALLOC(KMP_CPU_SET_SIZE_LIMIT);
|
||||
gCode = pthread_getaffinity_np(pthread_self(), KMP_CPU_SET_SIZE_LIMIT, reinterpret_cast<cpuset_t *>(buf));
|
||||
KA_TRACE(30, ("__kmp_affinity_determine_capable: "
|
||||
"initial getaffinity call returned %d errno = %d\n",
|
||||
gCode, errno));
|
||||
if (gCode == 0) {
|
||||
KMP_AFFINITY_ENABLE(KMP_CPU_SET_SIZE_LIMIT);
|
||||
KA_TRACE(10, ("__kmp_affinity_determine_capable: "
|
||||
"affinity supported (mask size %d)\n"<
|
||||
(int)__kmp_affin_mask_size));
|
||||
KMP_INTERNAL_FREE(buf);
|
||||
return;
|
||||
}
|
||||
#endif
|
||||
// save uncaught error code
|
||||
// int error = errno;
|
||||
KMP_INTERNAL_FREE(buf);
|
||||
@@ -827,13 +802,6 @@ void __kmp_create_worker(int gtid, kmp_info_t *th, size_t stack_size) {
|
||||
and also gives the user the stack space they requested for all threads */
|
||||
stack_size += gtid * __kmp_stkoffset * 2;
|
||||
|
||||
#if defined(__ANDROID__) && __ANDROID_API__ < 19
|
||||
// Round the stack size to a multiple of the page size. Older versions of
|
||||
// Android (until KitKat) would fail pthread_attr_setstacksize with EINVAL
|
||||
// if the stack size was not a multiple of the page size.
|
||||
stack_size = (stack_size + PAGE_SIZE - 1) & ~(PAGE_SIZE - 1);
|
||||
#endif
|
||||
|
||||
KA_TRACE(10, ("__kmp_create_worker: T#%d, default stacksize = %lu bytes, "
|
||||
"__kmp_stksize = %lu bytes, final stacksize = %lu bytes\n",
|
||||
gtid, KMP_DEFAULT_STKSIZE, __kmp_stksize, stack_size));
|
||||
@@ -2004,7 +1972,7 @@ int __kmp_is_address_mapped(void *addr) {
|
||||
int found = 0;
|
||||
int rc;
|
||||
|
||||
#if KMP_OS_LINUX || KMP_OS_HURD
|
||||
#if KMP_OS_LINUX || KMP_OS_FREEBSD || KMP_OS_HURD
|
||||
|
||||
/* On GNUish OSes, read the /proc/<pid>/maps pseudo-file to get all the address
|
||||
ranges mapped into the address space. */
|
||||
@@ -2042,44 +2010,6 @@ int __kmp_is_address_mapped(void *addr) {
|
||||
// Free resources.
|
||||
fclose(file);
|
||||
KMP_INTERNAL_FREE(name);
|
||||
#elif KMP_OS_FREEBSD
|
||||
char *buf;
|
||||
size_t lstsz;
|
||||
int mib[] = {CTL_KERN, KERN_PROC, KERN_PROC_VMMAP, getpid()};
|
||||
rc = sysctl(mib, 4, NULL, &lstsz, NULL, 0);
|
||||
if (rc < 0)
|
||||
return 0;
|
||||
// We pass from number of vm entry's semantic
|
||||
// to size of whole entry map list.
|
||||
lstsz = lstsz * 4 / 3;
|
||||
buf = reinterpret_cast<char *>(kmpc_malloc(lstsz));
|
||||
rc = sysctl(mib, 4, buf, &lstsz, NULL, 0);
|
||||
if (rc < 0) {
|
||||
kmpc_free(buf);
|
||||
return 0;
|
||||
}
|
||||
|
||||
char *lw = buf;
|
||||
char *up = buf + lstsz;
|
||||
|
||||
while (lw < up) {
|
||||
struct kinfo_vmentry *cur = reinterpret_cast<struct kinfo_vmentry *>(lw);
|
||||
size_t cursz = cur->kve_structsize;
|
||||
if (cursz == 0)
|
||||
break;
|
||||
void *start = reinterpret_cast<void *>(cur->kve_start);
|
||||
void *end = reinterpret_cast<void *>(cur->kve_end);
|
||||
// Readable/Writable addresses within current map entry
|
||||
if ((addr >= start) && (addr < end)) {
|
||||
if ((cur->kve_protection & KVME_PROT_READ) != 0 &&
|
||||
(cur->kve_protection & KVME_PROT_WRITE) != 0) {
|
||||
found = 1;
|
||||
break;
|
||||
}
|
||||
}
|
||||
lw += cursz;
|
||||
}
|
||||
kmpc_free(buf);
|
||||
|
||||
#elif KMP_OS_DARWIN
|
||||
|
||||
@@ -2401,8 +2331,7 @@ finish: // Clean up and exit.
|
||||
#endif // USE_LOAD_BALANCE
|
||||
|
||||
#if !(KMP_ARCH_X86 || KMP_ARCH_X86_64 || KMP_MIC || \
|
||||
((KMP_OS_LINUX || KMP_OS_DARWIN) && KMP_ARCH_AARCH64) || \
|
||||
KMP_ARCH_PPC64 || KMP_ARCH_RISCV64)
|
||||
((KMP_OS_LINUX || KMP_OS_DARWIN) && KMP_ARCH_AARCH64) || KMP_ARCH_PPC64)
|
||||
|
||||
// we really only need the case with 1 argument, because CLANG always build
|
||||
// a struct of pointers to shared variables referenced in the outlined function
|
||||
@@ -2486,6 +2415,10 @@ int __kmp_invoke_microtask(microtask_t pkfn, int gtid, int tid, int argc,
|
||||
break;
|
||||
}
|
||||
|
||||
#if OMPT_SUPPORT
|
||||
*exit_frame_ptr = 0;
|
||||
#endif
|
||||
|
||||
return 1;
|
||||
}
|
||||
|
||||
|
||||
@@ -168,26 +168,6 @@ ompt_label_##id:
|
||||
#define print_possible_return_addresses(addr) \
|
||||
printf("%" PRIu64 ": current_address=%p or %p\n", ompt_get_thread_data()->value, \
|
||||
((char *)addr) - 4, ((char *)addr) - 8)
|
||||
#elif KMP_ARCH_RISCV64
|
||||
#if __riscv_compressed
|
||||
// On RV64GC the C.NOP instruction is 2 byte long. In addition, the compiler
|
||||
// inserts a J instruction (targeting the successor basic block), which
|
||||
// accounts for another 4 bytes. Finally, an additional J instruction may
|
||||
// appear (adding 4 more bytes) when the C.NOP is referenced elsewhere (ie.
|
||||
// another branch).
|
||||
#define print_possible_return_addresses(addr) \
|
||||
printf("%" PRIu64 ": current_address=%p or %p\n", \
|
||||
ompt_get_thread_data()->value, ((char *)addr) - 6, ((char *)addr) - 10)
|
||||
#else
|
||||
// On RV64G the NOP instruction is 4 byte long. In addition, the compiler
|
||||
// inserts a J instruction (targeting the successor basic block), which
|
||||
// accounts for another 4 bytes. Finally, an additional J instruction may
|
||||
// appear (adding 4 more bytes) when the NOP is referenced elsewhere (ie.
|
||||
// another branch).
|
||||
#define print_possible_return_addresses(addr) \
|
||||
printf("%" PRIu64 ": current_address=%p or %p\n", \
|
||||
ompt_get_thread_data()->value, ((char *)addr) - 8, ((char *)addr) - 12)
|
||||
#endif
|
||||
#else
|
||||
#error Unsupported target architecture, cannot determine address offset!
|
||||
#endif
|
||||
@@ -492,8 +472,7 @@ on_ompt_callback_implicit_task(
|
||||
char buffer[2048];
|
||||
|
||||
format_task_type(flags, buffer);
|
||||
// Only check initial task not created by teams construct
|
||||
if (team_size == 1 && thread_num == 1 && parallel_data->ptr)
|
||||
if(parallel_data->ptr)
|
||||
printf("%s\n", "0: parallel_data initially not null");
|
||||
parallel_data->value = ompt_get_unique_id();
|
||||
printf("%" PRIu64 ": ompt_event_initial_task_begin: parallel_id=%" PRIu64 ", task_id=%" PRIu64 ", actual_parallelism=%" PRIu32 ", index=%" PRIu32 ", flags=%" PRIu32 "\n", ompt_get_thread_data()->value, parallel_data->value, task_data->value, team_size, thread_num, flags);
|
||||
@@ -504,12 +483,7 @@ on_ompt_callback_implicit_task(
|
||||
break;
|
||||
case ompt_scope_end:
|
||||
if(flags & ompt_task_initial){
|
||||
printf("%" PRIu64 ": ompt_event_initial_task_end: parallel_id=%" PRIu64
|
||||
", task_id=%" PRIu64 ", actual_parallelism=%" PRIu32
|
||||
", index=%" PRIu32 "\n",
|
||||
ompt_get_thread_data()->value,
|
||||
(parallel_data) ? parallel_data->value : 0, task_data->value,
|
||||
team_size, thread_num);
|
||||
printf("%" PRIu64 ": ompt_event_initial_task_end: parallel_id=%" PRIu64 ", task_id=%" PRIu64 ", team_size=%" PRIu32 ", thread_num=%" PRIu32 "\n", ompt_get_thread_data()->value, (parallel_data)?parallel_data->value:0, task_data->value, team_size, thread_num);
|
||||
} else {
|
||||
printf("%" PRIu64 ": ompt_event_implicit_task_end: parallel_id=%" PRIu64 ", task_id=%" PRIu64 ", team_size=%" PRIu32 ", thread_num=%" PRIu32 "\n", ompt_get_thread_data()->value, (parallel_data)?parallel_data->value:0, task_data->value, team_size, thread_num);
|
||||
}
|
||||
@@ -650,28 +624,23 @@ static void on_ompt_callback_parallel_begin(
|
||||
if(parallel_data->ptr)
|
||||
printf("0: parallel_data initially not null\n");
|
||||
parallel_data->value = ompt_get_unique_id();
|
||||
int invoker = flag & 0xF;
|
||||
const char *event = (flag & ompt_parallel_team) ? "parallel" : "teams";
|
||||
const char *size = (flag & ompt_parallel_team) ? "team_size" : "num_teams";
|
||||
printf("%" PRIu64 ": ompt_event_%s_begin: parent_task_id=%" PRIu64
|
||||
printf("%" PRIu64 ": ompt_event_parallel_begin: parent_task_id=%" PRIu64
|
||||
", parent_task_frame.exit=%p, parent_task_frame.reenter=%p, "
|
||||
"parallel_id=%" PRIu64 ", requested_%s=%" PRIu32
|
||||
"parallel_id=%" PRIu64 ", requested_team_size=%" PRIu32
|
||||
", codeptr_ra=%p, invoker=%d\n",
|
||||
ompt_get_thread_data()->value, event, encountering_task_data->value,
|
||||
ompt_get_thread_data()->value, encountering_task_data->value,
|
||||
encountering_task_frame->exit_frame.ptr,
|
||||
encountering_task_frame->enter_frame.ptr, parallel_data->value, size,
|
||||
requested_team_size, codeptr_ra, invoker);
|
||||
encountering_task_frame->enter_frame.ptr, parallel_data->value,
|
||||
requested_team_size, codeptr_ra, flag);
|
||||
}
|
||||
|
||||
static void on_ompt_callback_parallel_end(ompt_data_t *parallel_data,
|
||||
ompt_data_t *encountering_task_data,
|
||||
int flag, const void *codeptr_ra) {
|
||||
int invoker = flag & 0xF;
|
||||
const char *event = (flag & ompt_parallel_team) ? "parallel" : "teams";
|
||||
printf("%" PRIu64 ": ompt_event_%s_end: parallel_id=%" PRIu64
|
||||
printf("%" PRIu64 ": ompt_event_parallel_end: parallel_id=%" PRIu64
|
||||
", task_id=%" PRIu64 ", invoker=%d, codeptr_ra=%p\n",
|
||||
ompt_get_thread_data()->value, event, parallel_data->value,
|
||||
encountering_task_data->value, invoker, codeptr_ra);
|
||||
ompt_get_thread_data()->value, parallel_data->value,
|
||||
encountering_task_data->value, flag, codeptr_ra);
|
||||
}
|
||||
|
||||
static void
|
||||
|
||||
@@ -58,8 +58,7 @@ int main() {
|
||||
|
||||
|
||||
// CHECK: {{^}}[[MASTER_ID:[0-9]+]]: ompt_event_initial_task_begin: parallel_id={{[0-9]+}}
|
||||
// CHECK-SAME: task_id=[[PARENT_TASK_ID_1:[0-9]+]], actual_parallelism=1,
|
||||
// CHECK-SAME: index=1, flags=1
|
||||
// CHECK-SAME: task_id=[[PARENT_TASK_ID_1:[0-9]+]], actual_parallelism=1, index=1, flags=1
|
||||
|
||||
// CHECK: {{^}}[[MASTER_ID_1]]: ompt_event_parallel_begin:
|
||||
// CHECK-SAME: parent_task_id=[[PARENT_TASK_ID_1]]
|
||||
@@ -74,7 +73,7 @@ int main() {
|
||||
|
||||
// CHECK: {{^}}[[MASTER_ID_1]]: ompt_event_initial_task_end:
|
||||
// CHECK-SAME: parallel_id={{[0-9]+}}, task_id=[[PARENT_TASK_ID_1]],
|
||||
// CHECK-SAME: actual_parallelism=0, index=1
|
||||
// CHECK-SAME: team_size=0, thread_num=1
|
||||
|
||||
// CHECK: {{^}}[[MASTER_ID_1]]: ompt_event_thread_end:
|
||||
// CHECK-SAME: thread_id=[[MASTER_ID_1]]
|
||||
@@ -84,8 +83,7 @@ int main() {
|
||||
// CHECK-SAME: thread_type=ompt_thread_initial=1, thread_id=[[MASTER_ID_2]]
|
||||
|
||||
// CHECK: {{^}}[[MASTER_ID:[0-9]+]]: ompt_event_initial_task_begin: parallel_id={{[0-9]+}}
|
||||
// CHECK-SAME: task_id=[[PARENT_TASK_ID_2:[0-9]+]], actual_parallelism=1,
|
||||
// CHECK-SAME: index=1, flags=1
|
||||
// CHECK-SAME: task_id=[[PARENT_TASK_ID_2:[0-9]+]], actual_parallelism=1, index=1, flags=1
|
||||
|
||||
// CHECK: {{^}}[[MASTER_ID_2]]: ompt_event_parallel_begin:
|
||||
// CHECK-SAME: parent_task_id=[[PARENT_TASK_ID_2]]
|
||||
@@ -101,7 +99,7 @@ int main() {
|
||||
|
||||
// CHECK: {{^}}[[MASTER_ID_2]]: ompt_event_initial_task_end:
|
||||
// CHECK-SAME: parallel_id={{[0-9]+}}, task_id=[[PARENT_TASK_ID_2]],
|
||||
// CHECK-SAME: actual_parallelism=0, index=1
|
||||
// CHECK-SAME: team_size=0, thread_num=1
|
||||
|
||||
// CHECK: {{^}}[[MASTER_ID_2]]: ompt_event_thread_end:
|
||||
// CHECK-SAME: thread_id=[[MASTER_ID_2]]
|
||||
|
||||
@@ -64,9 +64,6 @@ int main() {
|
||||
// THREADS: {{^}}[[MASTER_ID]]: ompt_event_barrier_begin
|
||||
// THREADS-SAME: parallel_id=[[PARALLEL_ID]], task_id=[[IMPLICIT_TASK_ID]]
|
||||
// THREADS-SAME: codeptr_ra=[[RETURN_ADDRESS]]{{[0-f][0-f]}}
|
||||
// THREADS: {{^}}[[MASTER_ID]]: task level 0
|
||||
// THREADS-SAME: parallel_id=[[PARALLEL_ID]], task_id=[[IMPLICIT_TASK_ID]]
|
||||
// THREADS-SAME: exit_frame=[[NULL]], reenter_frame=[[NULL]]
|
||||
// THREADS: {{^}}[[MASTER_ID]]: ompt_event_barrier_end
|
||||
// parallel_id is 0 because the region ended in the barrier!
|
||||
// THREADS-SAME: parallel_id=0, task_id=[[IMPLICIT_TASK_ID]]
|
||||
|
||||
@@ -1,81 +0,0 @@
|
||||
// RUN: %libomp-compile-and-run | %sort-threads | FileCheck %s
|
||||
// REQUIRES: ompt
|
||||
// UNSUPPORTED: gcc
|
||||
#include "callback.h"
|
||||
|
||||
int main() {
|
||||
#pragma omp target teams num_teams(1) thread_limit(2)
|
||||
#pragma omp parallel num_threads(2)
|
||||
{ printf("In teams\n"); }
|
||||
return 0;
|
||||
}
|
||||
|
||||
// CHECK: 0: NULL_POINTER=[[NULL:.*$]]
|
||||
|
||||
// CHECK-NOT: 0: parallel_data initially not null
|
||||
// CHECK-NOT: 0: task_data initially not null
|
||||
// CHECK-NOT: 0: thread_data initially not null
|
||||
|
||||
// CHECK: {{^}}[[MASTER:[0-9]+]]: ompt_event_initial_task_begin:
|
||||
// CHECK-SAME: task_id=[[INIT_TASK:[0-9]+]], {{.*}}, index=1
|
||||
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_teams_begin:
|
||||
// CHECK-SAME: parent_task_id=[[INIT_TASK]]
|
||||
// CHECK-SAME: {{.*}} requested_num_teams=1
|
||||
// CHECK-SAME: {{.*}} invoker=[[TEAMS_FLAGS:[0-9]+]]
|
||||
|
||||
//
|
||||
// team 0/thread 0
|
||||
//
|
||||
// initial task in the teams construct
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_initial_task_begin:
|
||||
// CHECK-SAME: task_id=[[INIT_TASK_0:[0-9]+]], actual_parallelism=1, index=0
|
||||
|
||||
// parallel region forked by runtime
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_parallel_begin:
|
||||
// CHECK-SAME: {{.*}} parent_task_id=[[INIT_TASK_0]]
|
||||
// CHECK-SAME: {{.*}} parallel_id=[[PAR_0:[0-9]+]]
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_implicit_task_begin:
|
||||
// CHECK-SAME: {{.*}} parallel_id=[[PAR_0]], task_id=[[IMPL_TASK_0:[0-9]+]]
|
||||
|
||||
// user parallel region
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_parallel_begin:
|
||||
// CHECK-SAME: {{.*}} parent_task_id=[[IMPL_TASK_0]]
|
||||
// CHECK-SAME: {{.*}} parallel_id=[[PAR_00:[0-9]+]]
|
||||
// CHECK-SAME: {{.*}} requested_team_size=2
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_implicit_task_begin:
|
||||
// CHECK-SAME: {{.*}} parallel_id=[[PAR_00]], task_id=[[IMPL_TASK_00:[0-9]+]]
|
||||
// CHECK-SAME: {{.*}} team_size=2, thread_num=0
|
||||
//
|
||||
// barrier event is here
|
||||
//
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_implicit_task_end:
|
||||
// CHECK-SAME: {{.*}} parallel_id={{[0-9]+}}, task_id=[[IMPL_TASK_00]]
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_parallel_end:
|
||||
// CHECK-SAME: {{.*}} parallel_id=[[PAR_00]], task_id=[[IMPL_TASK_0]]
|
||||
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_implicit_task_end:
|
||||
// CHECK-SAME: {{.*}} parallel_id={{[0-9]+}}, task_id=[[IMPL_TASK_0]]
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_parallel_end:
|
||||
// CHECK-SAME: {{.*}} parallel_id=[[PAR_0]], task_id=[[INIT_TASK_0]]
|
||||
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_initial_task_end:
|
||||
// CHECK-SAME: task_id=[[INIT_TASK_0]], actual_parallelism=0, index=0
|
||||
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_teams_end:
|
||||
// CHECK-SAME: {{.*}} task_id=[[INIT_TASK]], invoker=[[TEAMS_FLAGS]]
|
||||
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_initial_task_end:
|
||||
// CHECK-SAME: task_id=[[INIT_TASK]], {{.*}}, index=1
|
||||
|
||||
//
|
||||
// team 0/thread 1
|
||||
//
|
||||
// CHECK: {{^}}[[WORKER:[0-9]+]]: ompt_event_implicit_task_begin:
|
||||
// CHECK-SAME: {{.*}} parallel_id=[[PAR_00]], task_id=[[IMPL_TASK_01:[0-9]+]]
|
||||
// CHECK-SAME: {{.*}} team_size=2, thread_num=1
|
||||
//
|
||||
// barrier event is here
|
||||
//
|
||||
// CHECK: {{^}}[[WORKER]]: ompt_event_implicit_task_end:
|
||||
// CHECK-SAME: {{.*}} parallel_id={{[0-9]+}}, task_id=[[IMPL_TASK_01]]
|
||||
@@ -1,89 +0,0 @@
|
||||
// RUN: %libomp-compile-and-run | %sort-threads | FileCheck %s
|
||||
// REQUIRES: ompt
|
||||
// UNSUPPORTED: gcc
|
||||
#include "callback.h"
|
||||
|
||||
int main() {
|
||||
#pragma omp target teams num_teams(2) thread_limit(1)
|
||||
#pragma omp parallel num_threads(1)
|
||||
{ printf("In teams parallel\n"); }
|
||||
return 0;
|
||||
}
|
||||
|
||||
// CHECK: 0: NULL_POINTER=[[NULL:.*$]]
|
||||
|
||||
// CHECK-NOT: 0: parallel_data initially not null
|
||||
// CHECK-NOT: 0: task_data initially not null
|
||||
// CHECK-NOT: 0: thread_data initially not null
|
||||
|
||||
// CHECK: {{^}}[[MASTER_0:[0-9]+]]: ompt_event_initial_task_begin:
|
||||
// CHECK-SAME: task_id=[[INIT_TASK:[0-9]+]], {{.*}}, index=1
|
||||
|
||||
// CHECK: {{^}}[[MASTER_0]]: ompt_event_teams_begin:
|
||||
// CHECK-SAME: parent_task_id=[[INIT_TASK]]
|
||||
// CHECK-SAME: {{.*}} requested_num_teams=2
|
||||
// CHECK-SAME: {{.*}} invoker=[[TEAMS_FLAGS:[0-9]+]]
|
||||
|
||||
//
|
||||
// team 0
|
||||
//
|
||||
// initial task in the teams construct
|
||||
// CHECK: {{^}}[[MASTER_0]]: ompt_event_initial_task_begin:
|
||||
// CHECK-SAME: task_id=[[INIT_TASK_0:[0-9]+]], actual_parallelism=2, index=0
|
||||
|
||||
// parallel region forked by runtime
|
||||
// CHECK: {{^}}[[MASTER_0]]: ompt_event_parallel_begin:
|
||||
// CHECK-SAME: {{.*}} parent_task_id=[[INIT_TASK_0]]
|
||||
// CHECK-SAME: {{.*}} parallel_id=[[PAR_0:[0-9]+]]
|
||||
// CHECK: {{^}}[[MASTER_0]]: ompt_event_implicit_task_begin:
|
||||
// CHECK-SAME: {{.*}} parallel_id=[[PAR_0]], task_id=[[IMPL_TASK_0:[0-9]+]]
|
||||
|
||||
// user parallel region
|
||||
// CHECK: {{^}}[[MASTER_0]]: ompt_event_parallel_begin:
|
||||
// CHECK-SAME: {{.*}} parent_task_id=[[IMPL_TASK_0]]
|
||||
// CHECK-SAME: {{.*}} parallel_id=[[PAR_00:[0-9]+]]
|
||||
// CHECK: {{^}}[[MASTER_0]]: ompt_event_parallel_end:
|
||||
// CHECK-SAME: {{.*}} parallel_id=[[PAR_00]], task_id=[[IMPL_TASK_0]]
|
||||
|
||||
// CHECK: {{^}}[[MASTER_0]]: ompt_event_implicit_task_end:
|
||||
// CHECK-SAME: {{.*}} parallel_id={{[0-9]+}}, task_id=[[IMPL_TASK_0]]
|
||||
// CHECK: {{^}}[[MASTER_0]]: ompt_event_parallel_end:
|
||||
// CHECK-SAME: {{.*}} parallel_id=[[PAR_0]], task_id=[[INIT_TASK_0]]
|
||||
|
||||
// CHECK: {{^}}[[MASTER_0]]: ompt_event_initial_task_end:
|
||||
// CHECK-SAME: task_id=[[INIT_TASK_0]], actual_parallelism=0, index=0
|
||||
|
||||
// CHECK: {{^}}[[MASTER_0]]: ompt_event_teams_end:
|
||||
// CHECK-SAME: {{.*}} task_id=[[INIT_TASK]], invoker=[[TEAMS_FLAGS]]
|
||||
|
||||
// CHECK: {{^}}[[MASTER_0]]: ompt_event_initial_task_end:
|
||||
// CHECK-SAME: task_id=[[INIT_TASK]], {{.*}}, index=1
|
||||
|
||||
//
|
||||
// team 1
|
||||
//
|
||||
// initial task in the teams construct
|
||||
// CHECK: {{^}}[[MASTER_1:[0-9]+]]: ompt_event_initial_task_begin:
|
||||
// CHECK-SAME: task_id=[[INIT_TASK_1:[0-9]+]], actual_parallelism=2, index=1
|
||||
|
||||
// parallel region forked by runtime
|
||||
// CHECK: {{^}}[[MASTER_1]]: ompt_event_parallel_begin:
|
||||
// CHECK-SAME: {{.*}} parent_task_id=[[INIT_TASK_1]]
|
||||
// CHECK-SAME: {{.*}} parallel_id=[[PAR_ID_1:[0-9]+]]
|
||||
// CHECK: {{^}}[[MASTER_1]]: ompt_event_implicit_task_begin:
|
||||
// CHECK-SAME: {{.*}} parallel_id=[[PAR_ID_1]], task_id=[[IMPL_TASK_1:[0-9]+]]
|
||||
|
||||
// user parallel region
|
||||
// CHECK: {{^}}[[MASTER_1]]: ompt_event_parallel_begin:
|
||||
// CHECK-SAME: {{.*}} parent_task_id=[[IMPL_TASK_1]]
|
||||
// CHECK-SAME: {{.*}} parallel_id=[[PAR_ID_11:[0-9]+]]
|
||||
// CHECK: {{^}}[[MASTER_1]]: ompt_event_parallel_end:
|
||||
// CHECK-SAME: {{.*}} parallel_id=[[PAR_ID_11]], task_id=[[IMPL_TASK_1]]
|
||||
|
||||
// CHECK: {{^}}[[MASTER_1]]: ompt_event_implicit_task_end:
|
||||
// CHECK-SAME: {{.*}} parallel_id={{[0-9]+}}, task_id=[[IMPL_TASK_1]]
|
||||
// CHECK: {{^}}[[MASTER_1]]: ompt_event_parallel_end:
|
||||
// CHECK-SAME: {{.*}} parallel_id=[[PAR_ID_1]], task_id=[[INIT_TASK_1]]
|
||||
|
||||
// CHECK: {{^}}[[MASTER_1]]: ompt_event_initial_task_end:
|
||||
// CHECK-SAME: task_id=[[INIT_TASK_1]], actual_parallelism=0, index=1
|
||||
@@ -1,62 +0,0 @@
|
||||
// RUN: %libomp-compile-and-run | FileCheck %s
|
||||
// REQUIRES: ompt
|
||||
// UNSUPPORTED: gcc
|
||||
#include "callback.h"
|
||||
|
||||
int main() {
|
||||
#pragma omp target teams num_teams(1) thread_limit(1)
|
||||
#pragma omp parallel num_threads(1)
|
||||
{ printf("In teams\n"); }
|
||||
return 0;
|
||||
}
|
||||
|
||||
// CHECK: 0: NULL_POINTER=[[NULL:.*$]]
|
||||
|
||||
// CHECK-NOT: 0: parallel_data initially not null
|
||||
// CHECK-NOT: 0: task_data initially not null
|
||||
// CHECK-NOT: 0: thread_data initially not null
|
||||
|
||||
// CHECK: {{^}}[[MASTER:[0-9]+]]: ompt_event_initial_task_begin:
|
||||
// CHECK-SAME: task_id=[[INIT_TASK:[0-9]+]], {{.*}}, index=1
|
||||
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_teams_begin:
|
||||
// CHECK-SAME: parent_task_id=[[INIT_TASK]]
|
||||
// CHECK-SAME: {{.*}} requested_num_teams=1
|
||||
// CHECK-SAME: {{.*}} invoker=[[TEAMS_FLAGS:[0-9]+]]
|
||||
|
||||
// initial task in the teams construct starts
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_initial_task_begin:
|
||||
// CHECK-SAME: task_id=[[INIT_TASK_0:[0-9]+]], actual_parallelism=1, index=0
|
||||
|
||||
// parallel region forked by runtime
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_parallel_begin:
|
||||
// CHECK-SAME: {{.*}} parent_task_id=[[INIT_TASK_0]]
|
||||
// CHECK-SAME: {{.*}} parallel_id=[[PAR_0:[0-9]+]]
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_implicit_task_begin:
|
||||
// CHECK-SAME: {{.*}} parallel_id=[[PAR_0]], task_id=[[IMPL_TASK_0:[0-9]+]]
|
||||
|
||||
// user parallel region
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_parallel_begin:
|
||||
// CHECK-SAME: {{.*}} parent_task_id=[[IMPL_TASK_0]]
|
||||
// CHECK-SAME: {{.*}} parallel_id=[[PAR_00:[0-9]+]]
|
||||
// CHECK-SAME: {{.*}} requested_team_size=1
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_implicit_task_begin:
|
||||
// CHECK-SAME: {{.*}} parallel_id=[[PAR_00]], task_id=[[IMPL_TASK_00:[0-9]+]]
|
||||
// CHECK-SAME: {{.*}} team_size=1, thread_num=0
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_implicit_task_end:
|
||||
// CHECK-SAME: {{.*}} parallel_id={{[0-9]+}}, task_id=[[IMPL_TASK_00]]
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_parallel_end:
|
||||
// CHECK-SAME: {{.*}} parallel_id=[[PAR_00]], task_id=[[IMPL_TASK_0]]
|
||||
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_parallel_end:
|
||||
// CHECK-SAME: {{.*}} parallel_id=[[PAR_0]], task_id=[[INIT_TASK_0]]
|
||||
|
||||
// initial task in the teams construct ends
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_initial_task_end:
|
||||
// CHECK-SAME: task_id=[[INIT_TASK_0]], actual_parallelism=0, index=0
|
||||
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_teams_end:
|
||||
// CHECK-SAME: {{.*}} task_id=[[INIT_TASK]], invoker=[[TEAMS_FLAGS]]
|
||||
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_initial_task_end:
|
||||
// CHECK-SAME: task_id=[[INIT_TASK]], {{.*}}, index=1
|
||||
@@ -1,49 +0,0 @@
|
||||
// RUN: %libomp-compile-and-run | FileCheck %s
|
||||
// REQUIRES: ompt
|
||||
// UNSUPPORTED: gcc
|
||||
#include "callback.h"
|
||||
|
||||
int main() {
|
||||
#pragma omp target teams num_teams(1) thread_limit(1)
|
||||
{ printf("In teams\n"); }
|
||||
return 0;
|
||||
}
|
||||
|
||||
// CHECK: 0: NULL_POINTER=[[NULL:.*$]]
|
||||
|
||||
// CHECK-NOT: 0: parallel_data initially not null
|
||||
// CHECK-NOT: 0: task_data initially not null
|
||||
// CHECK-NOT: 0: thread_data initially not null
|
||||
|
||||
// CHECK: {{^}}[[MASTER:[0-9]+]]: ompt_event_initial_task_begin:
|
||||
// CHECK-SAME: task_id=[[INIT_TASK:[0-9]+]], {{.*}}, index=1
|
||||
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_teams_begin:
|
||||
// CHECK-SAME: parent_task_id=[[INIT_TASK]]
|
||||
// CHECK-SAME: {{.*}} requested_num_teams=1
|
||||
// CHECK-SAME: {{.*}} invoker=[[TEAMS_FLAGS:[0-9]+]]
|
||||
|
||||
// initial task in the teams construct starts
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_initial_task_begin:
|
||||
// CHECK-SAME: task_id=[[INIT_TASK_0:[0-9]+]], actual_parallelism=1, index=0
|
||||
|
||||
// parallel region forked by runtime
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_parallel_begin:
|
||||
// CHECK-SAME: {{.*}} parent_task_id=[[INIT_TASK_0]]
|
||||
// CHECK-SAME: {{.*}} parallel_id=[[PAR_0:[0-9]+]]
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_implicit_task_begin:
|
||||
// CHECK-SAME: {{.*}} parallel_id=[[PAR_0]], task_id=[[IMPL_TASK_0:[0-9]+]]
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_implicit_task_end:
|
||||
// CHECK-SAME: {{.*}} parallel_id={{[0-9]+}}, task_id=[[IMPL_TASK_0]]
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_parallel_end:
|
||||
// CHECK-SAME: {{.*}} parallel_id=[[PAR_0]], task_id=[[INIT_TASK_0]]
|
||||
|
||||
// initial task in the teams construct ends
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_initial_task_end:
|
||||
// CHECK-SAME: task_id=[[INIT_TASK_0]], actual_parallelism=0, index=0
|
||||
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_teams_end:
|
||||
// CHECK-SAME: {{.*}} task_id=[[INIT_TASK]], invoker=[[TEAMS_FLAGS]]
|
||||
|
||||
// CHECK: {{^}}[[MASTER]]: ompt_event_initial_task_end:
|
||||
// CHECK-SAME: task_id=[[INIT_TASK]], {{.*}}, index=1
|
||||
@@ -29,6 +29,7 @@ int main()
|
||||
{
|
||||
int i;
|
||||
int block = 1;
|
||||
int tid;
|
||||
int throttling = strcmp(getenv("KMP_ENABLE_TASK_THROTTLING"), "1") == 0;
|
||||
int enqueued = 0;
|
||||
int failed = -1;
|
||||
@@ -40,7 +41,6 @@ int main()
|
||||
enqueued++;
|
||||
#pragma omp task
|
||||
{
|
||||
int tid;
|
||||
tid = omp_get_thread_num();
|
||||
if (tid == 0) {
|
||||
// As soon as the master thread starts executing task we should unlock
|
||||
|
||||
@@ -1,38 +0,0 @@
|
||||
// RUN: %libomp-compile && env KMP_ENABLE_TASK_THROTTLING=0 %libomp-run
|
||||
|
||||
#include<omp.h>
|
||||
#include<stdlib.h>
|
||||
#include<string.h>
|
||||
|
||||
// The first hashtable static size is 997
|
||||
#define NUM_DEPS 4000
|
||||
|
||||
|
||||
int main()
|
||||
{
|
||||
int *deps = calloc(NUM_DEPS, sizeof(int));
|
||||
int i;
|
||||
int failed = 0;
|
||||
|
||||
#pragma omp parallel
|
||||
#pragma omp master
|
||||
{
|
||||
for (i = 0; i < NUM_DEPS; i++) {
|
||||
#pragma omp task firstprivate(i) depend(inout: deps[i])
|
||||
{
|
||||
deps[i] = 1;
|
||||
}
|
||||
#pragma omp task firstprivate(i) depend(inout: deps[i])
|
||||
{
|
||||
deps[i] = 2;
|
||||
}
|
||||
}
|
||||
}
|
||||
|
||||
for (i = 0; i < NUM_DEPS; i++) {
|
||||
if (deps[i] != 2)
|
||||
failed++;
|
||||
}
|
||||
|
||||
return failed;
|
||||
}
|
||||
@@ -131,7 +131,7 @@ sub get_deps_readelf($) {
|
||||
# Parse body.
|
||||
while ( $i < @bulk ) {
|
||||
my $line = $bulk[ $i ];
|
||||
if ( $line !~ m{^\s*0x[0-9a-f]+\s+\(?([_A-Z0-9]+)\)?\s+(.*)\s*$}i ) {
|
||||
if ( $line !~ m{^\s*0x[0-9a-f]+\s+\(([_A-Z0-9]+)\)\s+(.*)\s*$}i ) {
|
||||
parse_error( $tool, @bulk, $i );
|
||||
}; # if
|
||||
my ( $type, $value ) = ( $1, $2 );
|
||||
|
||||
@@ -61,8 +61,6 @@ sub canon_arch($) {
|
||||
$arch = "mips64";
|
||||
} elsif ( $arch =~ m{\Amips} ) {
|
||||
$arch = "mips";
|
||||
} elsif ( $arch =~ m{\Ariscv64} ) {
|
||||
$arch = "riscv64";
|
||||
} else {
|
||||
$arch = undef;
|
||||
}; # if
|
||||
@@ -96,7 +94,6 @@ sub canon_mic_arch($) {
|
||||
"mic" => "Intel(R) Many Integrated Core Architecture",
|
||||
"mips" => "MIPS",
|
||||
"mips64" => "MIPS64",
|
||||
"riscv64" => "RISC-V (64-bit)",
|
||||
);
|
||||
|
||||
sub legal_arch($) {
|
||||
@@ -223,8 +220,6 @@ sub target_options() {
|
||||
$_host_arch = "mips64";
|
||||
} elsif ( $hardware_platform eq "mips" ) {
|
||||
$_host_arch = "mips";
|
||||
} elsif ( $hardware_platform eq "riscv64" ) {
|
||||
$_host_arch = "riscv64";
|
||||
} else {
|
||||
die "Unsupported host hardware platform: \"$hardware_platform\"; stopped";
|
||||
}; # if
|
||||
@@ -414,7 +409,7 @@ the script assumes host architecture is target one.
|
||||
|
||||
Input string is an architecture name to canonize. The function recognizes many variants, for example:
|
||||
C<32e>, C<Intel64>, C<Intel(R) 64>, etc. Returned string is a canononized architecture name,
|
||||
one of: C<32>, C<32e>, C<64>, C<arm>, C<ppc64le>, C<ppc64>, C<mic>, C<mips>, C<mips64>, C<riscv64> or C<undef> is input string is not recognized.
|
||||
one of: C<32>, C<32e>, C<64>, C<arm>, C<ppc64le>, C<ppc64>, C<mic>, C<mips>, C<mips64>, or C<undef> is input string is not recognized.
|
||||
|
||||
=item B<legal_arch( $arch )>
|
||||
|
||||
|
||||
@@ -156,8 +156,6 @@ if ( 0 ) {
|
||||
$values{ hardware_platform } = "mips64";
|
||||
} elsif ( $values{ machine } =~ m{\Amips\z} ) {
|
||||
$values{ hardware_platform } = "mips";
|
||||
} elsif ( $values{ machine } =~ m{\Ariscv64\z} ) {
|
||||
$values{ hardware_platform } = "riscv64";
|
||||
} else {
|
||||
die "Unsupported machine (\"$values{ machine }\") returned by POSIX::uname(); stopped";
|
||||
}; # if
|
||||
|
||||
@@ -53,7 +53,6 @@ Architectures Supported
|
||||
* IBM(R) Power architecture (big endian)
|
||||
* IBM(R) Power architecture (little endian)
|
||||
* MIPS and MIPS64 architectures
|
||||
* RISC-V 64 bit architecture
|
||||
|
||||
Supported RTL Build Configurations
|
||||
==================================
|
||||
|
||||
@@ -134,7 +134,6 @@
|
||||
the Intel compiler.
|
||||
</li>
|
||||
<li>MIPS and MIPS64</li>
|
||||
<li>RISC-V 64-bit</li>
|
||||
</ul>
|
||||
Ports to other architectures and operating systems are welcome.
|
||||
</p>
|
||||
|
||||
Reference in New Issue
Block a user