diff --git a/openmp/libomptarget/CMakeLists.txt b/openmp/libomptarget/CMakeLists.txt --- a/openmp/libomptarget/CMakeLists.txt +++ b/openmp/libomptarget/CMakeLists.txt @@ -79,6 +79,7 @@ add_subdirectory(plugins) add_subdirectory(deviceRTLs) add_subdirectory(DeviceRTL) +add_subdirectory(tools) # Add tests. add_subdirectory(test) diff --git a/openmp/libomptarget/include/omptarget.h b/openmp/libomptarget/include/omptarget.h --- a/openmp/libomptarget/include/omptarget.h +++ b/openmp/libomptarget/include/omptarget.h @@ -210,6 +210,9 @@ /// adds a target shared library to the target execution image void __tgt_register_lib(__tgt_bin_desc *desc); +/// Initialize all RTLs at once +void __tgt_init_all_rtls(); + /// removes a target shared library from the target execution image void __tgt_unregister_lib(__tgt_bin_desc *desc); @@ -333,6 +336,7 @@ void __tgt_set_info_flag(uint32_t); +int __tgt_print_device_info(int64_t device_id); #ifdef __cplusplus } #endif diff --git a/openmp/libomptarget/include/omptargetplugin.h b/openmp/libomptarget/include/omptargetplugin.h --- a/openmp/libomptarget/include/omptargetplugin.h +++ b/openmp/libomptarget/include/omptargetplugin.h @@ -142,6 +142,9 @@ // Set plugin's internal information flag externally. void __tgt_rtl_set_info_flag(uint32_t); +// Print the device information +void __tgt_rtl_print_device_info(int32_t ID); + #ifdef __cplusplus } #endif diff --git a/openmp/libomptarget/plugins/cuda/dynamic_cuda/cuda.h b/openmp/libomptarget/plugins/cuda/dynamic_cuda/cuda.h --- a/openmp/libomptarget/plugins/cuda/dynamic_cuda/cuda.h +++ b/openmp/libomptarget/plugins/cuda/dynamic_cuda/cuda.h @@ -72,6 +72,11 @@ CUresult cuDeviceGetCount(int *); CUresult cuFuncGetAttribute(int *, CUfunction_attribute, CUfunction); +// Device info +CUresult cuDeviceGetName(char *, int, CUdevice *); +CUresult cuDeviceTotalMem(size_t *, CUdevice *); +CUresult cuDriverGetVersion(int *); + CUresult cuGetErrorString(CUresult, const char **); CUresult cuInit(unsigned); CUresult cuLaunchKernel(CUfunction, unsigned, unsigned, unsigned, unsigned, diff --git a/openmp/libomptarget/plugins/cuda/dynamic_cuda/cuda.cpp b/openmp/libomptarget/plugins/cuda/dynamic_cuda/cuda.cpp --- a/openmp/libomptarget/plugins/cuda/dynamic_cuda/cuda.cpp +++ b/openmp/libomptarget/plugins/cuda/dynamic_cuda/cuda.cpp @@ -28,6 +28,11 @@ DLWRAP(cuDeviceGetCount, 1); DLWRAP(cuFuncGetAttribute, 3); +// Device info +DLWRAP(cuDeviceGetName, 3); +DLWRAP(cuDeviceTotalMem, 2); +DLWRAP(cuDriverGetVersion, 1); + DLWRAP(cuGetErrorString, 2); DLWRAP(cuLaunchKernel, 11); diff --git a/openmp/libomptarget/plugins/cuda/src/rtl.cpp b/openmp/libomptarget/plugins/cuda/src/rtl.cpp --- a/openmp/libomptarget/plugins/cuda/src/rtl.cpp +++ b/openmp/libomptarget/plugins/cuda/src/rtl.cpp @@ -16,9 +16,11 @@ #include #include #include +#include #include #include #include +#include #include "Debug.h" #include "omptargetplugin.h" @@ -61,6 +63,8 @@ } while (false) #endif // OMPTARGET_DEBUG +#define BOOL2TEXT(b) ((b) ? "Yes" : "No") + #include "elf_common.h" /// Keep entries table per device. @@ -1157,6 +1161,178 @@ } return (Err == CUDA_SUCCESS) ? OFFLOAD_SUCCESS : OFFLOAD_FAIL; } + + void printDeviceInfo(int32_t device_id) { + char TmpChar[1000]; + std::string TmpStr; + size_t TmpSt; + int TmpInt, TmpInt2, TmpInt3; + + CUdevice Device; + checkResult(cuDeviceGet(&Device, device_id), + "Error returned from cuCtxGetDevice\n"); + + cuDriverGetVersion(&TmpInt); + printf(" CUDA Driver Version: \t\t%d \n", TmpInt); + printf(" CUDA Device Number: \t\t%d \n", device_id); + checkResult(cuDeviceGetName(TmpChar, 1000, Device), + "Error returned from cuDeviceGetName\n"); + printf(" Device Name: \t\t\t%s \n", TmpChar); + checkResult(cuDeviceTotalMem(&TmpSt, Device), + "Error returned from cuDeviceTotalMem\n"); + printf(" Global Memory Size: \t\t%zu bytes \n", TmpSt); + checkResult(cuDeviceGetAttribute( + &TmpInt, CU_DEVICE_ATTRIBUTE_MULTIPROCESSOR_COUNT, Device), + "Error returned from cuDeviceGetAttribute\n"); + printf(" Number of Multiprocessors: \t\t%d \n", TmpInt); + checkResult( + cuDeviceGetAttribute(&TmpInt, CU_DEVICE_ATTRIBUTE_GPU_OVERLAP, Device), + "Error returned from cuDeviceGetAttribute\n"); + printf(" Concurrent Copy and Execution: \t%s \n", BOOL2TEXT(TmpInt)); + checkResult(cuDeviceGetAttribute( + &TmpInt, CU_DEVICE_ATTRIBUTE_TOTAL_CONSTANT_MEMORY, Device), + "Error returned from cuDeviceGetAttribute\n"); + printf(" Total Constant Memory: \t\t%d bytes\n", TmpInt); + checkResult( + cuDeviceGetAttribute( + &TmpInt, CU_DEVICE_ATTRIBUTE_MAX_SHARED_MEMORY_PER_BLOCK, Device), + "Error returned from cuDeviceGetAttribute\n"); + printf(" Max Shared Memory per Block: \t%d bytes \n", TmpInt); + checkResult(cuDeviceGetAttribute( + &TmpInt, CU_DEVICE_ATTRIBUTE_MAX_REGISTERS_PER_BLOCK, Device), + "Error returned from cuDeviceGetAttribute\n"); + printf(" Registers per Block: \t\t%d \n", TmpInt); + checkResult( + cuDeviceGetAttribute(&TmpInt, CU_DEVICE_ATTRIBUTE_WARP_SIZE, Device), + "Error returned from cuDeviceGetAttribute\n"); + printf(" Warp Size: \t\t\t\t%d Threads \n", TmpInt); + checkResult(cuDeviceGetAttribute( + &TmpInt, CU_DEVICE_ATTRIBUTE_MAX_THREADS_PER_BLOCK, Device), + "Error returned from cuDeviceGetAttribute\n"); + printf(" Maximum Threads per Block: \t\t%d \n", TmpInt); + checkResult(cuDeviceGetAttribute(&TmpInt, CU_DEVICE_ATTRIBUTE_MAX_BLOCK_DIM_X, + Device), + "Error returned from cuDeviceGetAttribute\n"); + checkResult(cuDeviceGetAttribute(&TmpInt2, + CU_DEVICE_ATTRIBUTE_MAX_BLOCK_DIM_Y, Device), + "Error returned from cuDeviceGetAttribute\n"); + checkResult(cuDeviceGetAttribute(&TmpInt3, + CU_DEVICE_ATTRIBUTE_MAX_BLOCK_DIM_Z, Device), + "Error returned from cuDeviceGetAttribute\n"); + printf(" Maximum Block Dimensions: \t\t%d, %d, %d \n", TmpInt, TmpInt2, + TmpInt3); + checkResult( + cuDeviceGetAttribute(&TmpInt, CU_DEVICE_ATTRIBUTE_MAX_GRID_DIM_X, Device), + "Error returned from cuDeviceGetAttribute\n"); + checkResult(cuDeviceGetAttribute(&TmpInt2, CU_DEVICE_ATTRIBUTE_MAX_GRID_DIM_Y, + Device), + "Error returned from cuDeviceGetAttribute\n"); + checkResult(cuDeviceGetAttribute(&TmpInt3, CU_DEVICE_ATTRIBUTE_MAX_GRID_DIM_Z, + Device), + "Error returned from cuDeviceGetAttribute\n"); + printf(" Maximum Grid Dimensions: \t\t%d x %d x %d \n", TmpInt, TmpInt2, + TmpInt3); + checkResult( + cuDeviceGetAttribute(&TmpInt, CU_DEVICE_ATTRIBUTE_MAX_PITCH, Device), + "Error returned from cuDeviceGetAttribute\n"); + printf(" Maximum Memory Pitch: \t\t%d bytes \n", TmpInt); + checkResult(cuDeviceGetAttribute( + &TmpInt, CU_DEVICE_ATTRIBUTE_TEXTURE_ALIGNMENT, Device), + "Error returned from cuDeviceGetAttribute\n"); + printf(" Texture Alignment: \t\t\t%d bytes \n", TmpInt); + checkResult( + cuDeviceGetAttribute(&TmpInt, CU_DEVICE_ATTRIBUTE_CLOCK_RATE, Device), + "Error returned from cuDeviceGetAttribute\n"); + printf(" Clock Rate: \t\t\t%d kHz\n", TmpInt); + checkResult(cuDeviceGetAttribute( + &TmpInt, CU_DEVICE_ATTRIBUTE_KERNEL_EXEC_TIMEOUT, Device), + "Error returned from cuDeviceGetAttribute\n"); + printf(" Execution Timeout: \t\t\t%s \n", BOOL2TEXT(TmpInt)); + checkResult( + cuDeviceGetAttribute(&TmpInt, CU_DEVICE_ATTRIBUTE_INTEGRATED, Device), + "Error returned from cuDeviceGetAttribute\n"); + printf(" Integrated Device: \t\t\t%s \n", BOOL2TEXT(TmpInt)); + checkResult(cuDeviceGetAttribute( + &TmpInt, CU_DEVICE_ATTRIBUTE_CAN_MAP_HOST_MEMORY, Device), + "Error returned from cuDeviceGetAttribute\n"); + printf(" Can Map Host Memory: \t\t%s \n", BOOL2TEXT(TmpInt)); + checkResult( + cuDeviceGetAttribute(&TmpInt, CU_DEVICE_ATTRIBUTE_COMPUTE_MODE, Device), + "Error returned from cuDeviceGetAttribute\n"); + if (TmpInt == CU_COMPUTEMODE_DEFAULT) + TmpStr = "DEFAULT"; + else if (TmpInt == CU_COMPUTEMODE_PROHIBITED) + TmpStr = "PROHIBITED"; + else if (TmpInt == CU_COMPUTEMODE_EXCLUSIVE_PROCESS) + TmpStr = "EXCLUSIVE PROCESS"; + else + TmpStr = "unknown"; + printf(" Compute Mode: \t\t\t%s \n", TmpStr.c_str()); + checkResult(cuDeviceGetAttribute( + &TmpInt, CU_DEVICE_ATTRIBUTE_CONCURRENT_KERNELS, Device), + "Error returned from cuDeviceGetAttribute\n"); + printf(" Concurrent Kernels: \t\t%s \n", BOOL2TEXT(TmpInt)); + checkResult( + cuDeviceGetAttribute(&TmpInt, CU_DEVICE_ATTRIBUTE_ECC_ENABLED, Device), + "Error returned from cuDeviceGetAttribute\n"); + printf(" ECC Enabled: \t\t\t%s \n", BOOL2TEXT(TmpInt)); + checkResult(cuDeviceGetAttribute( + &TmpInt, CU_DEVICE_ATTRIBUTE_MEMORY_CLOCK_RATE, Device), + "Error returned from cuDeviceGetAttribute\n"); + printf(" Memory Clock Rate: \t\t\t%d kHz\n", TmpInt); + checkResult(cuDeviceGetAttribute( + &TmpInt, CU_DEVICE_ATTRIBUTE_GLOBAL_MEMORY_BUS_WIDTH, Device), + "Error returned from cuDeviceGetAttribute\n"); + printf(" Memory Bus Width: \t\t\t%d bits\n", TmpInt); + checkResult( + cuDeviceGetAttribute(&TmpInt, CU_DEVICE_ATTRIBUTE_L2_CACHE_SIZE, Device), + "Error returned from cuDeviceGetAttribute\n"); + printf(" L2 Cache Size: \t\t\t%d bytes \n", TmpInt); + checkResult( + cuDeviceGetAttribute( + &TmpInt, CU_DEVICE_ATTRIBUTE_MAX_THREADS_PER_MULTIPROCESSOR, Device), + "Error returned from cuDeviceGetAttribute\n"); + printf(" Max Threads Per SMP: \t\t%d \n", TmpInt); + checkResult(cuDeviceGetAttribute( + &TmpInt, CU_DEVICE_ATTRIBUTE_ASYNC_ENGINE_COUNT, Device), + "Error returned from cuDeviceGetAttribute\n"); + printf(" Async Engines: \t\t\t%s (%d) \n", BOOL2TEXT(TmpInt), TmpInt); + checkResult(cuDeviceGetAttribute( + &TmpInt, CU_DEVICE_ATTRIBUTE_UNIFIED_ADDRESSING, Device), + "Error returned from cuDeviceGetAttribute\n"); + printf(" Unified Addressing: \t\t%s \n", BOOL2TEXT(TmpInt)); + checkResult( + cuDeviceGetAttribute(&TmpInt, CU_DEVICE_ATTRIBUTE_MANAGED_MEMORY, Device), + "Error returned from cuDeviceGetAttribute\n"); + printf(" Managed Memory: \t\t\t%s \n", BOOL2TEXT(TmpInt)); + checkResult( + cuDeviceGetAttribute( + &TmpInt, CU_DEVICE_ATTRIBUTE_CONCURRENT_MANAGED_ACCESS, Device), + "Error returned from cuDeviceGetAttribute\n"); + printf(" Concurrent Managed Memory: \t\t%s \n", BOOL2TEXT(TmpInt)); + checkResult( + cuDeviceGetAttribute( + &TmpInt, CU_DEVICE_ATTRIBUTE_COMPUTE_PREEMPTION_SUPPORTED, Device), + "Error returned from cuDeviceGetAttribute\n"); + printf(" Preemption Supported: \t\t%s \n", BOOL2TEXT(TmpInt)); + checkResult(cuDeviceGetAttribute( + &TmpInt, CU_DEVICE_ATTRIBUTE_COOPERATIVE_LAUNCH, Device), + "Error returned from cuDeviceGetAttribute\n"); + printf(" Cooperative Launch: \t\t%s \n", BOOL2TEXT(TmpInt)); + checkResult(cuDeviceGetAttribute(&TmpInt, CU_DEVICE_ATTRIBUTE_MULTI_GPU_BOARD, + Device), + "Error returned from cuDeviceGetAttribute\n"); + printf(" Multi-Device Boars: \t\t%s \n", BOOL2TEXT(TmpInt)); + checkResult(cuDeviceGetAttribute(&TmpInt, + CU_DEVICE_ATTRIBUTE_COMPUTE_CAPABILITY_MAJOR, + Device), + "Error returned from cuDeviceGetAttribute\n"); + checkResult(cuDeviceGetAttribute(&TmpInt2, + CU_DEVICE_ATTRIBUTE_COMPUTE_CAPABILITY_MINOR, + Device), + "Error returned from cuDeviceGetAttribute\n"); + printf(" Compute Capabilities: \t\t%d%d \n", TmpInt, TmpInt2); + } }; DeviceRTLTy DeviceRTL; @@ -1357,6 +1533,11 @@ InfoLevel.store(NewInfoLevel); } +void __tgt_rtl_print_device_info(int32_t device_id) { + assert(DeviceRTL.isValidDeviceId(device_id) && "device_id is invalid"); + DeviceRTL.printDeviceInfo(device_id); +} + #ifdef __cplusplus } #endif diff --git a/openmp/libomptarget/plugins/exports b/openmp/libomptarget/plugins/exports --- a/openmp/libomptarget/plugins/exports +++ b/openmp/libomptarget/plugins/exports @@ -23,6 +23,7 @@ __tgt_rtl_unregister_lib; __tgt_rtl_supports_empty_images; __tgt_rtl_set_info_flag; + __tgt_rtl_print_device_info; local: *; }; diff --git a/openmp/libomptarget/src/device.h b/openmp/libomptarget/src/device.h --- a/openmp/libomptarget/src/device.h +++ b/openmp/libomptarget/src/device.h @@ -275,6 +275,10 @@ /// OFFLOAD_SUCCESS/OFFLOAD_FAIL when succeeds/fails. int32_t synchronize(AsyncInfoTy &AsyncInfo); + /// Calls the corresponding print in the \p RTLDEVID + /// device RTL to obtain the information of the specific device. + bool printDeviceInfo(int32_t RTLDevID); + private: // Call to RTL void init(); // To be called only via DeviceTy::initOnce() diff --git a/openmp/libomptarget/src/device.cpp b/openmp/libomptarget/src/device.cpp --- a/openmp/libomptarget/src/device.cpp +++ b/openmp/libomptarget/src/device.cpp @@ -511,6 +511,14 @@ TgtOffsets, TgtVarsSize, AsyncInfo); } +// Run region on device +bool DeviceTy::printDeviceInfo(int32_t RTLDevId) { + if (!RTL->print_device_info) + return false; + RTL->print_device_info(RTLDevId); + return true; +} + // Run team region on device. int32_t DeviceTy::runTeamRegion(void *TgtEntryPtr, void **TgtVarsPtr, ptrdiff_t *TgtOffsets, int32_t TgtVarsSize, diff --git a/openmp/libomptarget/src/exports b/openmp/libomptarget/src/exports --- a/openmp/libomptarget/src/exports +++ b/openmp/libomptarget/src/exports @@ -3,6 +3,7 @@ __tgt_register_requires; __tgt_register_lib; __tgt_unregister_lib; + __tgt_init_all_rtls; __tgt_target_data_begin; __tgt_target_data_end; __tgt_target_data_update; @@ -40,6 +41,7 @@ llvm_omp_target_alloc_shared; llvm_omp_target_alloc_device; __tgt_set_info_flag; + __tgt_print_device_info; local: *; }; diff --git a/openmp/libomptarget/src/interface.cpp b/openmp/libomptarget/src/interface.cpp --- a/openmp/libomptarget/src/interface.cpp +++ b/openmp/libomptarget/src/interface.cpp @@ -43,6 +43,10 @@ PM->RTLs.RegisterLib(desc); } +//////////////////////////////////////////////////////////////////////////////// +/// Initialize all available devices without registering any image +EXTERN void __tgt_init_all_rtls() { PM->RTLs.initAllRTLs(); } + //////////////////////////////////////////////////////////////////////////////// /// unloads a target shared library EXTERN void __tgt_unregister_lib(__tgt_bin_desc *desc) { @@ -466,3 +470,8 @@ R.set_info_flag(NewInfoLevel); } } + +EXTERN int __tgt_print_device_info(int64_t device_id) { + return PM->Devices[device_id].printDeviceInfo( + PM->Devices[device_id].RTLDeviceID); +} diff --git a/openmp/libomptarget/src/rtl.h b/openmp/libomptarget/src/rtl.h --- a/openmp/libomptarget/src/rtl.h +++ b/openmp/libomptarget/src/rtl.h @@ -55,6 +55,7 @@ typedef int64_t(synchronize_ty)(int32_t, __tgt_async_info *); typedef int32_t (*register_lib_ty)(__tgt_bin_desc *); typedef int32_t(supports_empty_images_ty)(); + typedef void(print_device_info_ty)(int32_t); typedef void(set_info_flag_ty)(uint32_t); int32_t Idx = -1; // RTL index, index is the number of devices @@ -93,6 +94,7 @@ register_lib_ty unregister_lib = nullptr; supports_empty_images_ty *supports_empty_images = nullptr; set_info_flag_ty *set_info_flag = nullptr; + print_device_info_ty *print_device_info = nullptr; // Are there images associated with this RTL. bool isUsed = false; @@ -119,6 +121,12 @@ // Register the clauses of the requires directive. void RegisterRequires(int64_t flags); + // Initialize RTL if it has not been initialized + void initRTLonce(RTLInfoTy &RTL); + + // Initialize all RTLs + void initAllRTLs(); + // Register a shared library with all (compatible) RTLs. void RegisterLib(__tgt_bin_desc *desc); diff --git a/openmp/libomptarget/src/rtl.cpp b/openmp/libomptarget/src/rtl.cpp --- a/openmp/libomptarget/src/rtl.cpp +++ b/openmp/libomptarget/src/rtl.cpp @@ -177,6 +177,8 @@ dlsym(dynlib_handle, "__tgt_rtl_supports_empty_images"); *((void **)&R.set_info_flag) = dlsym(dynlib_handle, "__tgt_rtl_set_info_flag"); + *((void **)&R.print_device_info) = + dlsym(dynlib_handle, "__tgt_rtl_print_device_info"); } DP("RTLs loaded!\n"); @@ -288,6 +290,38 @@ flags, RequiresFlags); } +void RTLsTy::initRTLonce(RTLInfoTy &R) { + // If this RTL is not already in use, initialize it. + if (!R.isUsed && R.NumberOfDevices != 0) { + // Initialize the device information for the RTL we are about to use. + DeviceTy device(&R); + size_t Start = PM->Devices.size(); + PM->Devices.resize(Start + R.NumberOfDevices, device); + for (int32_t device_id = 0; device_id < R.NumberOfDevices; device_id++) { + // global device ID + PM->Devices[Start + device_id].DeviceID = Start + device_id; + // RTL local device ID + PM->Devices[Start + device_id].RTLDeviceID = device_id; + } + + // Initialize the index of this RTL and save it in the used RTLs. + R.Idx = (UsedRTLs.empty()) + ? 0 + : UsedRTLs.back()->Idx + UsedRTLs.back()->NumberOfDevices; + assert((size_t)R.Idx == Start && + "RTL index should equal the number of devices used so far."); + R.isUsed = true; + UsedRTLs.push_back(&R); + + DP("RTL " DPxMOD " has index %d!\n", DPxPTR(R.LibraryHandler), R.Idx); + } +} + +void RTLsTy::initAllRTLs() { + for (auto &R : AllRTLs) + initRTLonce(R); +} + void RTLsTy::RegisterLib(__tgt_bin_desc *desc) { PM->RTLsMtx.lock(); // Register the images with the RTLs that understand them, if any. @@ -295,7 +329,7 @@ // Obtain the image. __tgt_device_image *img = &desc->DeviceImages[i]; - RTLInfoTy *FoundRTL = NULL; + RTLInfoTy *FoundRTL = nullptr; // Scan the RTLs that have associated images until we find one that supports // the current image. @@ -309,31 +343,7 @@ DP("Image " DPxMOD " is compatible with RTL %s!\n", DPxPTR(img->ImageStart), R.RTLName.c_str()); - // If this RTL is not already in use, initialize it. - if (!R.isUsed) { - // Initialize the device information for the RTL we are about to use. - DeviceTy device(&R); - size_t Start = PM->Devices.size(); - PM->Devices.resize(Start + R.NumberOfDevices, device); - for (int32_t device_id = 0; device_id < R.NumberOfDevices; - device_id++) { - // global device ID - PM->Devices[Start + device_id].DeviceID = Start + device_id; - // RTL local device ID - PM->Devices[Start + device_id].RTLDeviceID = device_id; - } - - // Initialize the index of this RTL and save it in the used RTLs. - R.Idx = (UsedRTLs.empty()) - ? 0 - : UsedRTLs.back()->Idx + UsedRTLs.back()->NumberOfDevices; - assert((size_t)R.Idx == Start && - "RTL index should equal the number of devices used so far."); - R.isUsed = true; - UsedRTLs.push_back(&R); - - DP("RTL " DPxMOD " has index %d!\n", DPxPTR(R.LibraryHandler), R.Idx); - } + initRTLonce(R); // Initialize (if necessary) translation table for this library. PM->TrlTblMtx.lock(); diff --git a/openmp/libomptarget/tools/CMakeLists.txt b/openmp/libomptarget/tools/CMakeLists.txt new file mode 100644 --- /dev/null +++ b/openmp/libomptarget/tools/CMakeLists.txt @@ -0,0 +1,13 @@ +##===----------------------------------------------------------------------===## +# +# 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 +# +##===----------------------------------------------------------------------===## +# +# Adding omptarget tools +# +##===----------------------------------------------------------------------===## + +add_subdirectory(deviceinfo) diff --git a/openmp/libomptarget/tools/deviceinfo/CMakeLists.txt b/openmp/libomptarget/tools/deviceinfo/CMakeLists.txt new file mode 100644 --- /dev/null +++ b/openmp/libomptarget/tools/deviceinfo/CMakeLists.txt @@ -0,0 +1,24 @@ +##===----------------------------------------------------------------------===## +# +# 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 +# +##===----------------------------------------------------------------------===## +# +# Build llvm-omp-device-info tool +# +##===----------------------------------------------------------------------===## + +libomptarget_say("Building the llvm-omp-device-info tool") +libomptarget_say("llvm-omp-device-info using plugins ${LIBOMPTARGET_TESTED_PLUGINS}") + +add_llvm_tool(llvm-omp-device-info llvm-omp-device-info.cpp) + +llvm_update_compile_flags(llvm-omp-device-info) + +target_link_libraries(llvm-omp-device-info PRIVATE + omp + omptarget + ${LIBOMPTARGET_TESTED_PLUGINS} +) diff --git a/openmp/libomptarget/tools/deviceinfo/llvm-omp-device-info.cpp b/openmp/libomptarget/tools/deviceinfo/llvm-omp-device-info.cpp new file mode 100644 --- /dev/null +++ b/openmp/libomptarget/tools/deviceinfo/llvm-omp-device-info.cpp @@ -0,0 +1,31 @@ +//===- llvm-omp-device-info.cpp - Obtain device info as seen from OpenMP --===// +// +// 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 +// +//===----------------------------------------------------------------------===// +// +// This is a command line utility that, by using Libomptarget, and the device +// plugins, list devices information as seen from the OpenMP Runtime. +// +//===----------------------------------------------------------------------===// + +#include "omptarget.h" +#include + +int main(int argc, char **argv) { + __tgt_bin_desc EmptyDesc = {0, nullptr, nullptr, nullptr}; + __tgt_register_lib(&EmptyDesc); + __tgt_init_all_rtls(); + + for (int Dev = 0; Dev < omp_get_num_devices(); Dev++) { + printf("Device (%d):\n", Dev); + if (!__tgt_print_device_info(Dev)) + printf(" print_device_info not implemented\n"); + printf("\n"); + } + + __tgt_unregister_lib(&EmptyDesc); + return 0; +}