Skip to content

[Offload] Move /openmp/libomptarget to /offload #75125

New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Merged
merged 2 commits into from
Apr 22, 2024

Conversation

jdoerfert
Copy link
Member

@jdoerfert jdoerfert commented Dec 12, 2023

In a nutshell, this moves our libomptarget code to create the offload subproject.

With this commit, users need to enable the new LLVM/Offload subproject as a runtime in their cmake configuration.
No further changes are expected for downstream code.

Tests and other components still depend on OpenMP and have also not been renamed. The results below are for a build in which OpenMP and Offload are enabled runtimes. In addition to the pure git mv, I needed to adjust some CMake files. Nothing is intended to change semantics.

ninja check-offload

Works with the X86 and AMDGPU offload tests

ninja check-openmp

Still works but doesn't build offload tests anymore.

ls install/lib

Shows all expected libraries, incl.

  • libomptarget.devicertl.a
  • libomptarget-nvptx-sm_90.bc
  • libomptarget.rtl.amdgpu.so -> libomptarget.rtl.amdgpu.so.18git
  • libomptarget.so -> libomptarget.so.18git

Fixes: #75124

@jdoerfert jdoerfert added openmp:libomptarget OpenMP offload runtime offload labels Dec 12, 2023
@jdoerfert jdoerfert self-assigned this Dec 12, 2023
@jdoerfert jdoerfert requested a review from a team as a code owner December 12, 2023 01:32
@llvmbot llvmbot added cmake Build system in general and CMake in particular backend:AMDGPU openmp:libomp OpenMP host runtime labels Dec 12, 2023
@llvmbot
Copy link
Member

llvmbot commented Dec 12, 2023

@llvm/pr-subscribers-backend-amdgpu

Author: Johannes Doerfert (jdoerfert)

Changes

** This is still WIP - Testing is needed **

In a nutshell, this moves our libomptarget code to create the offload subproject. For now, we allow LLVM/Offload only as runtime. Tests and other components still depend on OpenMP and have also not been renamed. The results below are for a build in which OpenMP and Offload are enabled runtimes. In addition to the pure git mv, I needed to adjust some CMake files. Nothing is intended to change semantics but some of the things likely broke other build configurations. Testers are needed.

ninja -C build/llvm/runtimes/runtimes-bins check-libomptarget

Works with the X86 and AMDGPU offload tests

ninja -C /build/llvm/ check-openmp

Still works but doesn't build offload tests anymore.

ls install/lib

Shows all expected libraries, incl.

  • libomptarget.devicertl.a
  • libomptarget-nvptx-sm_90.bc
  • libomptarget.rtl.amdgpu.so -> libomptarget.rtl.amdgpu.so.18git
  • libomptarget.so -> libomptarget.so.18git

Fixes: #75124


Patch is 92.57 KiB, truncated to 20.00 KiB below, full version: https://github.com/llvm/llvm-project/pull/75125.diff

324 Files Affected:

  • (modified) llvm/CMakeLists.txt (+1-1)
  • (renamed) offload/CMakeLists.txt (+93-9)
  • (renamed) offload/DeviceRTL/CMakeLists.txt (+2-2)
  • (renamed) offload/DeviceRTL/include/Allocator.h ()
  • (renamed) offload/DeviceRTL/include/Configuration.h ()
  • (renamed) offload/DeviceRTL/include/Debug.h ()
  • (renamed) offload/DeviceRTL/include/Interface.h ()
  • (renamed) offload/DeviceRTL/include/LibC.h ()
  • (renamed) offload/DeviceRTL/include/Mapping.h ()
  • (renamed) offload/DeviceRTL/include/State.h ()
  • (renamed) offload/DeviceRTL/include/Synchronization.h ()
  • (renamed) offload/DeviceRTL/include/Types.h ()
  • (renamed) offload/DeviceRTL/include/Utils.h ()
  • (renamed) offload/DeviceRTL/include/generated_microtask_cases.gen ()
  • (renamed) offload/DeviceRTL/src/Allocator.cpp ()
  • (renamed) offload/DeviceRTL/src/Configuration.cpp ()
  • (renamed) offload/DeviceRTL/src/Debug.cpp ()
  • (renamed) offload/DeviceRTL/src/Kernel.cpp ()
  • (renamed) offload/DeviceRTL/src/LibC.cpp ()
  • (renamed) offload/DeviceRTL/src/Mapping.cpp ()
  • (renamed) offload/DeviceRTL/src/Misc.cpp ()
  • (renamed) offload/DeviceRTL/src/Parallelism.cpp ()
  • (renamed) offload/DeviceRTL/src/Reduction.cpp ()
  • (renamed) offload/DeviceRTL/src/State.cpp ()
  • (renamed) offload/DeviceRTL/src/Stub.cpp ()
  • (renamed) offload/DeviceRTL/src/Synchronization.cpp ()
  • (renamed) offload/DeviceRTL/src/Tasking.cpp ()
  • (renamed) offload/DeviceRTL/src/Utils.cpp ()
  • (renamed) offload/DeviceRTL/src/Workshare.cpp ()
  • (renamed) offload/DeviceRTL/src/exports ()
  • (renamed) offload/README.txt ()
  • (renamed) offload/cmake/Modules/LibomptargetGetDependencies.cmake ()
  • (renamed) offload/cmake/Modules/LibomptargetUtils.cmake ()
  • (renamed) offload/docs/declare_target_indirect.md ()
  • (renamed) offload/include/DeviceImage.h ()
  • (renamed) offload/include/ExclusiveAccess.h ()
  • (renamed) offload/include/OffloadEntry.h ()
  • (renamed) offload/include/OffloadPolicy.h ()
  • (renamed) offload/include/OpenMP/InternalTypes.h ()
  • (renamed) offload/include/OpenMP/InteropAPI.h ()
  • (renamed) offload/include/OpenMP/Mapping.h ()
  • (renamed) offload/include/OpenMP/OMPT/Callback.h ()
  • (renamed) offload/include/OpenMP/OMPT/Connector.h ()
  • (renamed) offload/include/OpenMP/OMPT/Interface.h ()
  • (renamed) offload/include/OpenMP/omp.h ()
  • (renamed) offload/include/PluginManager.h ()
  • (renamed) offload/include/Shared/APITypes.h ()
  • (renamed) offload/include/Shared/Debug.h ()
  • (renamed) offload/include/Shared/Environment.h ()
  • (renamed) offload/include/Shared/EnvironmentVar.h ()
  • (renamed) offload/include/Shared/PluginAPI.h ()
  • (renamed) offload/include/Shared/PluginAPI.inc ()
  • (renamed) offload/include/Shared/Profile.h ()
  • (renamed) offload/include/Shared/Requirements.h ()
  • (renamed) offload/include/Shared/SourceInfo.h ()
  • (renamed) offload/include/Shared/Utils.h ()
  • (renamed) offload/include/Utils/ExponentialBackoff.h ()
  • (renamed) offload/include/device.h ()
  • (renamed) offload/include/omptarget.h ()
  • (renamed) offload/include/rtl.h ()
  • (renamed) offload/plugins-nextgen/CMakeLists.txt (+1-1)
  • (renamed) offload/plugins-nextgen/aarch64/CMakeLists.txt ()
  • (renamed) offload/plugins-nextgen/amdgpu/CMakeLists.txt (+1-1)
  • (renamed) offload/plugins-nextgen/amdgpu/dynamic_hsa/hsa.cpp ()
  • (renamed) offload/plugins-nextgen/amdgpu/dynamic_hsa/hsa.h ()
  • (renamed) offload/plugins-nextgen/amdgpu/dynamic_hsa/hsa_ext_amd.h ()
  • (renamed) offload/plugins-nextgen/amdgpu/src/rtl.cpp ()
  • (renamed) offload/plugins-nextgen/amdgpu/utils/UtilitiesRTL.h ()
  • (renamed) offload/plugins-nextgen/common/CMakeLists.txt ()
  • (renamed) offload/plugins-nextgen/common/OMPT/CMakeLists.txt ()
  • (renamed) offload/plugins-nextgen/common/OMPT/OmptCallback.cpp ()
  • (renamed) offload/plugins-nextgen/common/include/DLWrap.h ()
  • (renamed) offload/plugins-nextgen/common/include/GlobalHandler.h ()
  • (renamed) offload/plugins-nextgen/common/include/JIT.h ()
  • (renamed) offload/plugins-nextgen/common/include/MemoryManager.h ()
  • (renamed) offload/plugins-nextgen/common/include/PluginInterface.h ()
  • (renamed) offload/plugins-nextgen/common/include/RPC.h ()
  • (renamed) offload/plugins-nextgen/common/src/GlobalHandler.cpp ()
  • (renamed) offload/plugins-nextgen/common/src/JIT.cpp ()
  • (renamed) offload/plugins-nextgen/common/src/PluginInterface.cpp ()
  • (renamed) offload/plugins-nextgen/common/src/RPC.cpp ()
  • (renamed) offload/plugins-nextgen/common/src/Utils/ELF.cpp ()
  • (renamed) offload/plugins-nextgen/common/src/Utils/ELF.h ()
  • (renamed) offload/plugins-nextgen/cuda/CMakeLists.txt (+1-1)
  • (renamed) offload/plugins-nextgen/cuda/dynamic_cuda/cuda.cpp ()
  • (renamed) offload/plugins-nextgen/cuda/dynamic_cuda/cuda.h ()
  • (renamed) offload/plugins-nextgen/cuda/src/rtl.cpp ()
  • (renamed) offload/plugins-nextgen/exports ()
  • (renamed) offload/plugins-nextgen/generic-elf-64bit/src/rtl.cpp ()
  • (renamed) offload/plugins-nextgen/ppc64/CMakeLists.txt ()
  • (renamed) offload/plugins-nextgen/ppc64le/CMakeLists.txt ()
  • (renamed) offload/plugins-nextgen/s390x/CMakeLists.txt ()
  • (renamed) offload/plugins-nextgen/x86_64/CMakeLists.txt ()
  • (renamed) offload/src/CMakeLists.txt (+1-1)
  • (renamed) offload/src/DeviceImage.cpp ()
  • (renamed) offload/src/LegacyAPI.cpp ()
  • (renamed) offload/src/OffloadRTL.cpp ()
  • (renamed) offload/src/OpenMP/API.cpp ()
  • (renamed) offload/src/OpenMP/InteropAPI.cpp ()
  • (renamed) offload/src/OpenMP/Mapping.cpp ()
  • (renamed) offload/src/OpenMP/OMPT/Callback.cpp ()
  • (renamed) offload/src/PluginManager.cpp ()
  • (renamed) offload/src/device.cpp ()
  • (renamed) offload/src/exports ()
  • (renamed) offload/src/interface.cpp ()
  • (renamed) offload/src/omptarget.cpp ()
  • (renamed) offload/src/private.h ()
  • (renamed) offload/test/CMakeLists.txt ()
  • (renamed) offload/test/Inputs/basic_array.f90 ()
  • (renamed) offload/test/Inputs/declare_indirect_func.c ()
  • (renamed) offload/test/api/assert.c ()
  • (renamed) offload/test/api/is_initial_device.c ()
  • (renamed) offload/test/api/omp_device_managed_memory.c ()
  • (renamed) offload/test/api/omp_device_managed_memory_alloc.c ()
  • (renamed) offload/test/api/omp_device_memory.c ()
  • (renamed) offload/test/api/omp_dynamic_shared_memory.c ()
  • (renamed) offload/test/api/omp_dynamic_shared_memory_amdgpu.c ()
  • (renamed) offload/test/api/omp_dynamic_shared_memory_mixed.inc ()
  • (renamed) offload/test/api/omp_dynamic_shared_memory_mixed_amdgpu.c ()
  • (renamed) offload/test/api/omp_dynamic_shared_memory_mixed_nvptx.c ()
  • (renamed) offload/test/api/omp_env_vars.c ()
  • (renamed) offload/test/api/omp_get_device_num.c ()
  • (renamed) offload/test/api/omp_get_mapped_ptr.c ()
  • (renamed) offload/test/api/omp_get_num_devices.c ()
  • (renamed) offload/test/api/omp_get_num_devices_with_empty_target.c ()
  • (renamed) offload/test/api/omp_get_num_procs.c ()
  • (renamed) offload/test/api/omp_host_pinned_memory.c ()
  • (renamed) offload/test/api/omp_host_pinned_memory_alloc.c ()
  • (renamed) offload/test/api/omp_indirect_call.c ()
  • (renamed) offload/test/api/omp_target_memcpy_async1.c ()
  • (renamed) offload/test/api/omp_target_memcpy_async2.c ()
  • (renamed) offload/test/api/omp_target_memcpy_rect_async1.c ()
  • (renamed) offload/test/api/omp_target_memcpy_rect_async2.c ()
  • (renamed) offload/test/api/omp_target_memset.c ()
  • (renamed) offload/test/api/ompx_3d.c ()
  • (renamed) offload/test/api/ompx_3d.cpp ()
  • (renamed) offload/test/api/ompx_sync.c ()
  • (renamed) offload/test/api/ompx_sync.cpp ()
  • (renamed) offload/test/env/base_ptr_ref_count.c ()
  • (renamed) offload/test/env/omp_target_debug.c ()
  • (renamed) offload/test/jit/empty_kernel.inc ()
  • (renamed) offload/test/jit/empty_kernel_lvl1.c ()
  • (renamed) offload/test/jit/empty_kernel_lvl2.c ()
  • (renamed) offload/test/jit/type_punning.c ()
  • (renamed) offload/test/libc/assert.c ()
  • (renamed) offload/test/libc/fwrite.c ()
  • (renamed) offload/test/libc/global_ctor_dtor.cpp ()
  • (renamed) offload/test/libc/host_call.c ()
  • (renamed) offload/test/libc/malloc.c ()
  • (renamed) offload/test/libc/puts.c ()
  • (renamed) offload/test/lit.cfg ()
  • (renamed) offload/test/lit.site.cfg.in ()
  • (renamed) offload/test/mapping/alloc_fail.c ()
  • (renamed) offload/test/mapping/array_section_implicit_capture.c ()
  • (renamed) offload/test/mapping/array_section_use_device_ptr.c ()
  • (renamed) offload/test/mapping/data_absent_at_exit.c ()
  • (renamed) offload/test/mapping/data_member_ref.cpp ()
  • (renamed) offload/test/mapping/declare_mapper_api.cpp ()
  • (renamed) offload/test/mapping/declare_mapper_nested_default_mappers.cpp ()
  • (renamed) offload/test/mapping/declare_mapper_nested_default_mappers_array.cpp ()
  • (renamed) offload/test/mapping/declare_mapper_nested_default_mappers_array_subscript.cpp ()
  • (renamed) offload/test/mapping/declare_mapper_nested_default_mappers_complex_structure.cpp ()
  • (renamed) offload/test/mapping/declare_mapper_nested_default_mappers_ptr_subscript.cpp ()
  • (renamed) offload/test/mapping/declare_mapper_nested_default_mappers_var.cpp ()
  • (renamed) offload/test/mapping/declare_mapper_nested_mappers.cpp ()
  • (renamed) offload/test/mapping/declare_mapper_target.cpp ()
  • (renamed) offload/test/mapping/declare_mapper_target_data.cpp ()
  • (renamed) offload/test/mapping/declare_mapper_target_data_enter_exit.cpp ()
  • (renamed) offload/test/mapping/declare_mapper_target_update.cpp ()
  • (renamed) offload/test/mapping/delete_inf_refcount.c ()
  • (renamed) offload/test/mapping/device_ptr_update.c ()
  • (renamed) offload/test/mapping/firstprivate_aligned.cpp ()
  • (renamed) offload/test/mapping/has_device_addr.cpp ()
  • (renamed) offload/test/mapping/implicit_device_ptr.c ()
  • (renamed) offload/test/mapping/is_device_ptr.cpp ()
  • (renamed) offload/test/mapping/lambda_by_value.cpp ()
  • (renamed) offload/test/mapping/lambda_mapping.cpp ()
  • (renamed) offload/test/mapping/low_alignment.c ()
  • (renamed) offload/test/mapping/map_back_race.cpp ()
  • (renamed) offload/test/mapping/ompx_hold/omp_target_disassociate_ptr.c ()
  • (renamed) offload/test/mapping/ompx_hold/struct.c ()
  • (renamed) offload/test/mapping/ompx_hold/target-data.c ()
  • (renamed) offload/test/mapping/ompx_hold/target.c ()
  • (renamed) offload/test/mapping/padding_not_mapped.c ()
  • (renamed) offload/test/mapping/power_of_two_alignment.c ()
  • (renamed) offload/test/mapping/pr38704.c ()
  • (renamed) offload/test/mapping/prelock.cpp ()
  • (renamed) offload/test/mapping/present/target.c ()
  • (renamed) offload/test/mapping/present/target_array_extension.c ()
  • (renamed) offload/test/mapping/present/target_data.c ()
  • (renamed) offload/test/mapping/present/target_data_array_extension.c ()
  • (renamed) offload/test/mapping/present/target_data_at_exit.c ()
  • (renamed) offload/test/mapping/present/target_enter_data.c ()
  • (renamed) offload/test/mapping/present/target_exit_data_delete.c ()
  • (renamed) offload/test/mapping/present/target_exit_data_release.c ()
  • (renamed) offload/test/mapping/present/target_update.c ()
  • (renamed) offload/test/mapping/present/target_update_array_extension.c ()
  • (renamed) offload/test/mapping/present/unified_shared_memory.c ()
  • (renamed) offload/test/mapping/present/zero_length_array_section.c ()
  • (renamed) offload/test/mapping/present/zero_length_array_section_exit.c ()
  • (renamed) offload/test/mapping/private_mapping.c ()
  • (renamed) offload/test/mapping/ptr_and_obj_motion.c ()
  • (renamed) offload/test/mapping/reduction_implicit_map.cpp ()
  • (renamed) offload/test/mapping/target_data_array_extension_at_exit.c ()
  • (renamed) offload/test/mapping/target_derefence_array_pointrs.cpp ()
  • (renamed) offload/test/mapping/target_has_device_addr.c ()
  • (renamed) offload/test/mapping/target_implicit_partial_map.c ()
  • (renamed) offload/test/mapping/target_map_for_member_data.cpp ()
  • (renamed) offload/test/mapping/target_pointers_members_map.cpp ()
  • (renamed) offload/test/mapping/target_update_array_extension.c ()
  • (renamed) offload/test/mapping/target_use_device_addr.c ()
  • (renamed) offload/test/mapping/target_uses_allocator.c ()
  • (renamed) offload/test/mapping/target_wrong_use_device_addr.c ()
  • (renamed) offload/test/offloading/assert.cpp ()
  • (renamed) offload/test/offloading/atomic-compare-signedness.c ()
  • (renamed) offload/test/offloading/back2back_distribute.c ()
  • (renamed) offload/test/offloading/barrier_fence.c ()
  • (renamed) offload/test/offloading/bug47654.cpp ()
  • (renamed) offload/test/offloading/bug49021.cpp ()
  • (renamed) offload/test/offloading/bug49334.cpp ()
  • (renamed) offload/test/offloading/bug49779.cpp ()
  • (renamed) offload/test/offloading/bug50022.cpp ()
  • (renamed) offload/test/offloading/bug51781.c ()
  • (renamed) offload/test/offloading/bug51982.c ()
  • (renamed) offload/test/offloading/bug53727.cpp ()
  • (renamed) offload/test/offloading/bug64959.c ()
  • (renamed) offload/test/offloading/bug64959_compile_only.c ()
  • (renamed) offload/test/offloading/bug74582.c ()
  • (renamed) offload/test/offloading/complex_reduction.cpp ()
  • (renamed) offload/test/offloading/ctor_dtor.cpp ()
  • (renamed) offload/test/offloading/cuda_no_devices.c ()
  • (renamed) offload/test/offloading/d2d_memcpy.c ()
  • (renamed) offload/test/offloading/default_thread_limit.c ()
  • (renamed) offload/test/offloading/dynamic_module.c ()
  • (renamed) offload/test/offloading/dynamic_module_load.c ()
  • (renamed) offload/test/offloading/extern.c ()
  • (renamed) offload/test/offloading/fortran/basic-target-region-1D-array-section.f90 ()
  • (renamed) offload/test/offloading/fortran/basic-target-region-3D-array-section.f90 ()
  • (renamed) offload/test/offloading/fortran/basic-target-region-3D-array.f90 ()
  • (renamed) offload/test/offloading/fortran/basic_array.c ()
  • (renamed) offload/test/offloading/fortran/basic_target_region.f90 ()
  • (renamed) offload/test/offloading/fortran/constant-arr-index.f90 ()
  • (renamed) offload/test/offloading/fortran/declare-target-array-in-target-region.f90 ()
  • (renamed) offload/test/offloading/fortran/double-target-call-with-declare-target.f90 ()
  • (renamed) offload/test/offloading/fortran/failing/target_map_common_block1.f90 ()
  • (renamed) offload/test/offloading/fortran/target-region-implicit-array.f90 ()
  • (renamed) offload/test/offloading/fortran/target_map_common_block.f90 ()
  • (renamed) offload/test/offloading/fortran/target_map_common_block2.f90 ()
  • (renamed) offload/test/offloading/generic_multiple_parallel_regions.c ()
  • (renamed) offload/test/offloading/global_constructor.cpp ()
  • (renamed) offload/test/offloading/host_as_target.c ()
  • (renamed) offload/test/offloading/indirect_fp_mapping.c ()
  • (renamed) offload/test/offloading/info.c ()
  • (renamed) offload/test/offloading/interop.c ()
  • (renamed) offload/test/offloading/lone_target_exit_data.c ()
  • (renamed) offload/test/offloading/looptripcnt.c ()
  • (renamed) offload/test/offloading/malloc.c ()
  • (renamed) offload/test/offloading/malloc_parallel.c ()
  • (renamed) offload/test/offloading/mandatory_but_no_devices.c ()
  • (renamed) offload/test/offloading/memory_manager.cpp ()
  • (renamed) offload/test/offloading/multiple_reductions_simple.c ()
  • (renamed) offload/test/offloading/non_contiguous_update.cpp ()
  • (renamed) offload/test/offloading/offloading_success.c ()
  • (renamed) offload/test/offloading/offloading_success.cpp ()
  • (renamed) offload/test/offloading/ompx_coords.c ()
  • (renamed) offload/test/offloading/ompx_saxpy_mixed.c ()
  • (renamed) offload/test/offloading/parallel_offloading_map.cpp ()
  • (renamed) offload/test/offloading/parallel_target_teams_reduction.cpp ()
  • (renamed) offload/test/offloading/parallel_target_teams_reduction_max.cpp ()
  • (renamed) offload/test/offloading/parallel_target_teams_reduction_min.cpp ()
  • (renamed) offload/test/offloading/requires.c ()
  • (renamed) offload/test/offloading/shared_lib_fp_mapping.c ()
  • (renamed) offload/test/offloading/small_trip_count.c ()
  • (renamed) offload/test/offloading/small_trip_count_thread_limit.cpp ()
  • (renamed) offload/test/offloading/spmdization.c ()
  • (renamed) offload/test/offloading/static_linking.c ()
  • (renamed) offload/test/offloading/std_complex_arithmetic.cpp ()
  • (renamed) offload/test/offloading/target-teams-atomic.c ()
  • (renamed) offload/test/offloading/target-tile.c ()
  • (renamed) offload/test/offloading/target_constexpr_mapping.cpp ()
  • (renamed) offload/test/offloading/target_critical_region.cpp ()
  • (renamed) offload/test/offloading/target_depend_nowait.cpp ()
  • (renamed) offload/test/offloading/target_map_for_member_data.cpp ()
  • (renamed) offload/test/offloading/target_nowait_target.cpp ()
  • (renamed) offload/test/offloading/task_in_reduction_target.c ()
  • (renamed) offload/test/offloading/taskloop_offload_nowait.cpp ()
  • (renamed) offload/test/offloading/test_libc.cpp ()
  • (renamed) offload/test/offloading/thread_limit.c ()
  • (renamed) offload/test/offloading/thread_state_1.c ()
  • (renamed) offload/test/offloading/thread_state_2.c ()
  • (renamed) offload/test/offloading/weak.c ()
  • (renamed) offload/test/offloading/wtime.c ()
  • (renamed) offload/test/ompt/callbacks.h ()
  • (renamed) offload/test/ompt/register_both.h ()
  • (renamed) offload/test/ompt/register_emi.h ()
  • (renamed) offload/test/ompt/register_emi_map.h ()
  • (renamed) offload/test/ompt/register_no_device_init.h ()
  • (renamed) offload/test/ompt/register_non_emi.h ()
  • (renamed) offload/test/ompt/register_non_emi_map.h ()
  • (renamed) offload/test/ompt/register_wrong_return.h ()
  • (renamed) offload/test/ompt/veccopy.c ()
  • (renamed) offload/test/ompt/veccopy_data.c ()
  • (renamed) offload/test/ompt/veccopy_disallow_both.c ()
  • (renamed) offload/test/ompt/veccopy_emi.c ()
  • (renamed) offload/test/ompt/veccopy_emi_map.c ()
  • (renamed) offload/test/ompt/veccopy_map.c ()
  • (renamed) offload/test/ompt/veccopy_no_device_init.c ()
  • (renamed) offload/test/ompt/veccopy_wrong_return.c ()
  • (renamed) offload/test/unified_shared_memory/api.c ()
  • (renamed) offload/test/unified_shared_memory/associate_ptr.c ()
  • (renamed) offload/test/unified_shared_memory/close_enter_exit.c ()
  • (renamed) offload/test/unified_shared_memory/close_manual.c ()
  • (renamed) offload/test/unified_shared_memory/close_member.c ()
  • (renamed) offload/test/unified_shared_memory/close_modifier.c ()
  • (renamed) offload/test/unified_shared_memory/shared_update.c ()
  • (renamed) offload/tools/CMakeLists.txt ()
  • (renamed) offload/tools/deviceinfo/CMakeLists.txt ()
  • (renamed) offload/tools/deviceinfo/llvm-omp-device-info.cpp ()
  • (renamed) offload/tools/kernelreplay/CMakeLists.txt ()
  • (renamed) offload/tools/kernelreplay/llvm-omp-kernel-replay.cpp ()
  • (renamed) offload/utils/generate_microtask_cases.py ()
  • (modified) openmp/CMakeLists.txt (-24)
  • (modified) openmp/runtime/src/CMakeLists.txt (+7-1)
  • (modified) runtimes/CMakeLists.txt (+1-1)
diff --git a/llvm/CMakeLists.txt b/llvm/CMakeLists.txt
index 35be9bf16b5e1e..47dc1f493b12ee 100644
--- a/llvm/CMakeLists.txt
+++ b/llvm/CMakeLists.txt
@@ -156,7 +156,7 @@ endif()
 # As we migrate runtimes to using the bootstrapping build, the set of default runtimes
 # should grow as we remove those runtimes from LLVM_ENABLE_PROJECTS above.
 set(LLVM_DEFAULT_RUNTIMES "libcxx;libcxxabi;libunwind")
-set(LLVM_SUPPORTED_RUNTIMES "libc;libunwind;libcxxabi;pstl;libcxx;compiler-rt;openmp;llvm-libgcc")
+set(LLVM_SUPPORTED_RUNTIMES "libc;libunwind;libcxxabi;pstl;libcxx;compiler-rt;openmp;llvm-libgcc;offload")
 set(LLVM_ENABLE_RUNTIMES "" CACHE STRING
   "Semicolon-separated list of runtimes to build, or \"all\" (${LLVM_DEFAULT_RUNTIMES}). Supported runtimes are ${LLVM_SUPPORTED_RUNTIMES}.")
 if(LLVM_ENABLE_RUNTIMES STREQUAL "all")
diff --git a/openmp/libomptarget/CMakeLists.txt b/offload/CMakeLists.txt
similarity index 61%
rename from openmp/libomptarget/CMakeLists.txt
rename to offload/CMakeLists.txt
index 66925ccbe03054..8a2e4ffb5cc6e0 100644
--- a/openmp/libomptarget/CMakeLists.txt
+++ b/offload/CMakeLists.txt
@@ -10,12 +10,98 @@
 #
 ##===----------------------------------------------------------------------===##
 
-if("${CMAKE_SOURCE_DIR}" STREQUAL "${CMAKE_CURRENT_SOURCE_DIR}")
-  message(FATAL_ERROR "Direct configuration not supported, please use parent directory!")
+set(ENABLE_LIBOMPTARGET ON)
+# Currently libomptarget cannot be compiled on Windows or MacOS X.
+# Since the device plugins are only supported on Linux anyway,
+# there is no point in trying to compile libomptarget on other OSes.
+# 32-bit systems are not supported either.
+if (APPLE OR WIN32 OR NOT "cxx_std_17" IN_LIST CMAKE_CXX_COMPILE_FEATURES OR NOT CMAKE_SIZEOF_VOID_P EQUAL 8)
+  set(ENABLE_LIBOMPTARGET OFF)
 endif()
 
-# Add cmake directory to search for custom cmake functions.
-set(CMAKE_MODULE_PATH ${CMAKE_CURRENT_SOURCE_DIR}/cmake/Modules ${CMAKE_MODULE_PATH})
+option(OPENMP_ENABLE_LIBOMPTARGET "Enable building libomptarget for offloading."
+       ${ENABLE_LIBOMPTARGET})
+if (OPENMP_ENABLE_LIBOMPTARGET)
+  # Check that the library can actually be built.
+  if (APPLE OR WIN32)
+    message(FATAL_ERROR "libomptarget cannot be built on Windows and MacOS X!")
+  elseif (NOT "cxx_std_17" IN_LIST CMAKE_CXX_COMPILE_FEATURES)
+    message(FATAL_ERROR "Host compiler must support C++17 to build libomptarget!")
+  elseif (NOT CMAKE_SIZEOF_VOID_P EQUAL 8)
+    message(FATAL_ERROR "libomptarget on 32-bit systems are not supported!")
+  endif()
+endif()
+
+# TODO: Leftover from the move, could probably be just LLVM_LIBDIR_SUFFIX everywhere.
+set(OFFLOAD_INSTALL_LIBDIR "lib${LLVM_LIBDIR_SUFFIX}")
+
+set(LLVM_COMMON_CMAKE_UTILS ${CMAKE_CURRENT_SOURCE_DIR}/cmake)
+
+# Add path for custom modules
+list(INSERT CMAKE_MODULE_PATH 0
+  "${CMAKE_CURRENT_SOURCE_DIR}/cmake"
+  "${LLVM_COMMON_CMAKE_UTILS}/Modules"
+  )
+
+if (OPENMP_STANDALONE_BUILD)
+  # CMAKE_BUILD_TYPE was not set, default to Release.
+  if (NOT CMAKE_BUILD_TYPE)
+    set(CMAKE_BUILD_TYPE Release)
+  endif()
+
+  # Group common settings.
+  set(OPENMP_ENABLE_WERROR FALSE CACHE BOOL
+    "Enable -Werror flags to turn warnings into errors for supporting compilers.")
+  set(OPENMP_LIBDIR_SUFFIX "" CACHE STRING
+    "Suffix of lib installation directory, e.g. 64 => lib64")
+  # Do not use OPENMP_LIBDIR_SUFFIX directly, use OPENMP_INSTALL_LIBDIR.
+  set(OPENMP_INSTALL_LIBDIR "lib${OPENMP_LIBDIR_SUFFIX}")
+
+  # Group test settings.
+  set(OPENMP_TEST_C_COMPILER ${CMAKE_C_COMPILER} CACHE STRING
+    "C compiler to use for testing OpenMP runtime libraries.")
+  set(OPENMP_TEST_CXX_COMPILER ${CMAKE_CXX_COMPILER} CACHE STRING
+    "C++ compiler to use for testing OpenMP runtime libraries.")
+  set(OPENMP_TEST_Fortran_COMPILER ${CMAKE_Fortran_COMPILER} CACHE STRING
+    "FORTRAN compiler to use for testing OpenMP runtime libraries.")
+  set(OPENMP_LLVM_TOOLS_DIR "" CACHE PATH "Path to LLVM tools for testing.")
+
+  set(CMAKE_CXX_STANDARD 17 CACHE STRING "C++ standard to conform to")
+  set(CMAKE_CXX_STANDARD_REQUIRED NO)
+  set(CMAKE_CXX_EXTENSIONS NO)
+else()
+  set(OPENMP_ENABLE_WERROR ${LLVM_ENABLE_WERROR})
+  # If building in tree, we honor the same install suffix LLVM uses.
+  set(OPENMP_INSTALL_LIBDIR "lib${LLVM_LIBDIR_SUFFIX}")
+
+  if (NOT MSVC)
+    set(OPENMP_TEST_C_COMPILER ${LLVM_RUNTIME_OUTPUT_INTDIR}/clang)
+    set(OPENMP_TEST_CXX_COMPILER ${LLVM_RUNTIME_OUTPUT_INTDIR}/clang++)
+  else()
+    set(OPENMP_TEST_C_COMPILER ${LLVM_RUNTIME_OUTPUT_INTDIR}/clang.exe)
+    set(OPENMP_TEST_CXX_COMPILER ${LLVM_RUNTIME_OUTPUT_INTDIR}/clang++.exe)
+  endif()
+
+  # Check for flang
+  if (NOT MSVC)
+    set(OPENMP_TEST_Fortran_COMPILER ${LLVM_RUNTIME_OUTPUT_INTDIR}/flang-new)
+  else()
+    set(OPENMP_TEST_Fortran_COMPILER ${LLVM_RUNTIME_OUTPUT_INTDIR}/flang-new.exe)
+  endif()
+
+  # Set fortran test compiler if flang is found
+  if (EXISTS "${OPENMP_TEST_Fortran_COMPILER}")
+    message("Using local flang build at ${OPENMP_TEST_Fortran_COMPILER}")
+  else()
+    unset(OPENMP_TEST_Fortran_COMPILER)
+  endif()
+
+  # If not standalone, set CMAKE_CXX_STANDARD but don't set the global cache value,
+  # only set it locally for OpenMP.
+  set(CMAKE_CXX_STANDARD 17)
+  set(CMAKE_CXX_STANDARD_REQUIRED NO)
+  set(CMAKE_CXX_EXTENSIONS NO)
+endif()
 
 # Set the path of all resulting libraries to a unified location so that it can
 # be used for testing.
@@ -36,6 +122,9 @@ include(LibomptargetUtils)
 # Get dependencies for the different components of the project.
 include(LibomptargetGetDependencies)
 
+# Set up testing infrastructure.
+include(OpenMPTesting)
+
 # LLVM source tree is required at build time for libomptarget
 if (NOT LIBOMPTARGET_LLVM_INCLUDE_DIRS)
   message(FATAL_ERROR "Missing definition for LIBOMPTARGET_LLVM_INCLUDE_DIRS")
@@ -129,11 +218,6 @@ set(LIBOMPTARGET_INCLUDE_DIR ${CMAKE_CURRENT_SOURCE_DIR}/include)
 message(STATUS "OpenMP tools dir in libomptarget: ${LIBOMP_OMP_TOOLS_INCLUDE_DIR}")
 include_directories(${LIBOMP_OMP_TOOLS_INCLUDE_DIR})
 
-# Definitions for testing, for reuse when testing libomptarget-nvptx.
-set(LIBOMPTARGET_OPENMP_HEADER_FOLDER "${LIBOMP_INCLUDE_DIR}" CACHE STRING
-  "Path to folder containing omp.h")
-set(LIBOMPTARGET_OPENMP_HOST_RTL_FOLDER "${LIBOMP_LIBRARY_DIR}" CACHE STRING
-  "Path to folder containing libomp.so, and libLLVMSupport.so with profiling enabled")
 set(LIBOMPTARGET_LLVM_LIBRARY_DIR "${LLVM_LIBRARY_DIR}" CACHE STRING
   "Path to folder containing llvm library libomptarget.so")
 set(LIBOMPTARGET_LLVM_LIBRARY_INTDIR "${LIBOMPTARGET_INTDIR}" CACHE STRING
diff --git a/openmp/libomptarget/DeviceRTL/CMakeLists.txt b/offload/DeviceRTL/CMakeLists.txt
similarity index 99%
rename from openmp/libomptarget/DeviceRTL/CMakeLists.txt
rename to offload/DeviceRTL/CMakeLists.txt
index 1ce3e1e40a80ab..fa282603b2c114 100644
--- a/openmp/libomptarget/DeviceRTL/CMakeLists.txt
+++ b/offload/DeviceRTL/CMakeLists.txt
@@ -228,7 +228,7 @@ function(compileDeviceRTLLibrary target_cpu target_name target_triple)
   set_property(DIRECTORY APPEND PROPERTY ADDITIONAL_MAKE_CLEAN_FILES ${bclib_name} ${LIBOMPTARGET_LIBRARY_DIR}/${bclib_name})
 
   # Install bitcode library under the lib destination folder.
-  install(FILES ${CMAKE_CURRENT_BINARY_DIR}/${bclib_name} DESTINATION "${OPENMP_INSTALL_LIBDIR}")
+  install(FILES ${CMAKE_CURRENT_BINARY_DIR}/${bclib_name} DESTINATION "${OFFLOAD_INSTALL_LIBDIR}")
 
   set(target_feature "")
   if("${target_triple}" STREQUAL "nvptx64-nvidia-cuda")
@@ -307,4 +307,4 @@ set_target_properties(omptarget.devicertl PROPERTIES
 )
 target_link_libraries(omptarget.devicertl PRIVATE omptarget.devicertl.all_objs)
 
-install(TARGETS omptarget.devicertl ARCHIVE DESTINATION ${OPENMP_INSTALL_LIBDIR})
+install(TARGETS omptarget.devicertl ARCHIVE DESTINATION ${OFFLOAD_INSTALL_LIBDIR})
diff --git a/openmp/libomptarget/DeviceRTL/include/Allocator.h b/offload/DeviceRTL/include/Allocator.h
similarity index 100%
rename from openmp/libomptarget/DeviceRTL/include/Allocator.h
rename to offload/DeviceRTL/include/Allocator.h
diff --git a/openmp/libomptarget/DeviceRTL/include/Configuration.h b/offload/DeviceRTL/include/Configuration.h
similarity index 100%
rename from openmp/libomptarget/DeviceRTL/include/Configuration.h
rename to offload/DeviceRTL/include/Configuration.h
diff --git a/openmp/libomptarget/DeviceRTL/include/Debug.h b/offload/DeviceRTL/include/Debug.h
similarity index 100%
rename from openmp/libomptarget/DeviceRTL/include/Debug.h
rename to offload/DeviceRTL/include/Debug.h
diff --git a/openmp/libomptarget/DeviceRTL/include/Interface.h b/offload/DeviceRTL/include/Interface.h
similarity index 100%
rename from openmp/libomptarget/DeviceRTL/include/Interface.h
rename to offload/DeviceRTL/include/Interface.h
diff --git a/openmp/libomptarget/DeviceRTL/include/LibC.h b/offload/DeviceRTL/include/LibC.h
similarity index 100%
rename from openmp/libomptarget/DeviceRTL/include/LibC.h
rename to offload/DeviceRTL/include/LibC.h
diff --git a/openmp/libomptarget/DeviceRTL/include/Mapping.h b/offload/DeviceRTL/include/Mapping.h
similarity index 100%
rename from openmp/libomptarget/DeviceRTL/include/Mapping.h
rename to offload/DeviceRTL/include/Mapping.h
diff --git a/openmp/libomptarget/DeviceRTL/include/State.h b/offload/DeviceRTL/include/State.h
similarity index 100%
rename from openmp/libomptarget/DeviceRTL/include/State.h
rename to offload/DeviceRTL/include/State.h
diff --git a/openmp/libomptarget/DeviceRTL/include/Synchronization.h b/offload/DeviceRTL/include/Synchronization.h
similarity index 100%
rename from openmp/libomptarget/DeviceRTL/include/Synchronization.h
rename to offload/DeviceRTL/include/Synchronization.h
diff --git a/openmp/libomptarget/DeviceRTL/include/Types.h b/offload/DeviceRTL/include/Types.h
similarity index 100%
rename from openmp/libomptarget/DeviceRTL/include/Types.h
rename to offload/DeviceRTL/include/Types.h
diff --git a/openmp/libomptarget/DeviceRTL/include/Utils.h b/offload/DeviceRTL/include/Utils.h
similarity index 100%
rename from openmp/libomptarget/DeviceRTL/include/Utils.h
rename to offload/DeviceRTL/include/Utils.h
diff --git a/openmp/libomptarget/DeviceRTL/include/generated_microtask_cases.gen b/offload/DeviceRTL/include/generated_microtask_cases.gen
similarity index 100%
rename from openmp/libomptarget/DeviceRTL/include/generated_microtask_cases.gen
rename to offload/DeviceRTL/include/generated_microtask_cases.gen
diff --git a/openmp/libomptarget/DeviceRTL/src/Allocator.cpp b/offload/DeviceRTL/src/Allocator.cpp
similarity index 100%
rename from openmp/libomptarget/DeviceRTL/src/Allocator.cpp
rename to offload/DeviceRTL/src/Allocator.cpp
diff --git a/openmp/libomptarget/DeviceRTL/src/Configuration.cpp b/offload/DeviceRTL/src/Configuration.cpp
similarity index 100%
rename from openmp/libomptarget/DeviceRTL/src/Configuration.cpp
rename to offload/DeviceRTL/src/Configuration.cpp
diff --git a/openmp/libomptarget/DeviceRTL/src/Debug.cpp b/offload/DeviceRTL/src/Debug.cpp
similarity index 100%
rename from openmp/libomptarget/DeviceRTL/src/Debug.cpp
rename to offload/DeviceRTL/src/Debug.cpp
diff --git a/openmp/libomptarget/DeviceRTL/src/Kernel.cpp b/offload/DeviceRTL/src/Kernel.cpp
similarity index 100%
rename from openmp/libomptarget/DeviceRTL/src/Kernel.cpp
rename to offload/DeviceRTL/src/Kernel.cpp
diff --git a/openmp/libomptarget/DeviceRTL/src/LibC.cpp b/offload/DeviceRTL/src/LibC.cpp
similarity index 100%
rename from openmp/libomptarget/DeviceRTL/src/LibC.cpp
rename to offload/DeviceRTL/src/LibC.cpp
diff --git a/openmp/libomptarget/DeviceRTL/src/Mapping.cpp b/offload/DeviceRTL/src/Mapping.cpp
similarity index 100%
rename from openmp/libomptarget/DeviceRTL/src/Mapping.cpp
rename to offload/DeviceRTL/src/Mapping.cpp
diff --git a/openmp/libomptarget/DeviceRTL/src/Misc.cpp b/offload/DeviceRTL/src/Misc.cpp
similarity index 100%
rename from openmp/libomptarget/DeviceRTL/src/Misc.cpp
rename to offload/DeviceRTL/src/Misc.cpp
diff --git a/openmp/libomptarget/DeviceRTL/src/Parallelism.cpp b/offload/DeviceRTL/src/Parallelism.cpp
similarity index 100%
rename from openmp/libomptarget/DeviceRTL/src/Parallelism.cpp
rename to offload/DeviceRTL/src/Parallelism.cpp
diff --git a/openmp/libomptarget/DeviceRTL/src/Reduction.cpp b/offload/DeviceRTL/src/Reduction.cpp
similarity index 100%
rename from openmp/libomptarget/DeviceRTL/src/Reduction.cpp
rename to offload/DeviceRTL/src/Reduction.cpp
diff --git a/openmp/libomptarget/DeviceRTL/src/State.cpp b/offload/DeviceRTL/src/State.cpp
similarity index 100%
rename from openmp/libomptarget/DeviceRTL/src/State.cpp
rename to offload/DeviceRTL/src/State.cpp
diff --git a/openmp/libomptarget/DeviceRTL/src/Stub.cpp b/offload/DeviceRTL/src/Stub.cpp
similarity index 100%
rename from openmp/libomptarget/DeviceRTL/src/Stub.cpp
rename to offload/DeviceRTL/src/Stub.cpp
diff --git a/openmp/libomptarget/DeviceRTL/src/Synchronization.cpp b/offload/DeviceRTL/src/Synchronization.cpp
similarity index 100%
rename from openmp/libomptarget/DeviceRTL/src/Synchronization.cpp
rename to offload/DeviceRTL/src/Synchronization.cpp
diff --git a/openmp/libomptarget/DeviceRTL/src/Tasking.cpp b/offload/DeviceRTL/src/Tasking.cpp
similarity index 100%
rename from openmp/libomptarget/DeviceRTL/src/Tasking.cpp
rename to offload/DeviceRTL/src/Tasking.cpp
diff --git a/openmp/libomptarget/DeviceRTL/src/Utils.cpp b/offload/DeviceRTL/src/Utils.cpp
similarity index 100%
rename from openmp/libomptarget/DeviceRTL/src/Utils.cpp
rename to offload/DeviceRTL/src/Utils.cpp
diff --git a/openmp/libomptarget/DeviceRTL/src/Workshare.cpp b/offload/DeviceRTL/src/Workshare.cpp
similarity index 100%
rename from openmp/libomptarget/DeviceRTL/src/Workshare.cpp
rename to offload/DeviceRTL/src/Workshare.cpp
diff --git a/openmp/libomptarget/DeviceRTL/src/exports b/offload/DeviceRTL/src/exports
similarity index 100%
rename from openmp/libomptarget/DeviceRTL/src/exports
rename to offload/DeviceRTL/src/exports
diff --git a/openmp/libomptarget/README.txt b/offload/README.txt
similarity index 100%
rename from openmp/libomptarget/README.txt
rename to offload/README.txt
diff --git a/openmp/libomptarget/cmake/Modules/LibomptargetGetDependencies.cmake b/offload/cmake/Modules/LibomptargetGetDependencies.cmake
similarity index 100%
rename from openmp/libomptarget/cmake/Modules/LibomptargetGetDependencies.cmake
rename to offload/cmake/Modules/LibomptargetGetDependencies.cmake
diff --git a/openmp/libomptarget/cmake/Modules/LibomptargetUtils.cmake b/offload/cmake/Modules/LibomptargetUtils.cmake
similarity index 100%
rename from openmp/libomptarget/cmake/Modules/LibomptargetUtils.cmake
rename to offload/cmake/Modules/LibomptargetUtils.cmake
diff --git a/openmp/libomptarget/docs/declare_target_indirect.md b/offload/docs/declare_target_indirect.md
similarity index 100%
rename from openmp/libomptarget/docs/declare_target_indirect.md
rename to offload/docs/declare_target_indirect.md
diff --git a/openmp/libomptarget/include/DeviceImage.h b/offload/include/DeviceImage.h
similarity index 100%
rename from openmp/libomptarget/include/DeviceImage.h
rename to offload/include/DeviceImage.h
diff --git a/openmp/libomptarget/include/ExclusiveAccess.h b/offload/include/ExclusiveAccess.h
similarity index 100%
rename from openmp/libomptarget/include/ExclusiveAccess.h
rename to offload/include/ExclusiveAccess.h
diff --git a/openmp/libomptarget/include/OffloadEntry.h b/offload/include/OffloadEntry.h
similarity index 100%
rename from openmp/libomptarget/include/OffloadEntry.h
rename to offload/include/OffloadEntry.h
diff --git a/openmp/libomptarget/include/OffloadPolicy.h b/offload/include/OffloadPolicy.h
similarity index 100%
rename from openmp/libomptarget/include/OffloadPolicy.h
rename to offload/include/OffloadPolicy.h
diff --git a/openmp/libomptarget/include/OpenMP/InternalTypes.h b/offload/include/OpenMP/InternalTypes.h
similarity index 100%
rename from openmp/libomptarget/include/OpenMP/InternalTypes.h
rename to offload/include/OpenMP/InternalTypes.h
diff --git a/openmp/libomptarget/include/OpenMP/InteropAPI.h b/offload/include/OpenMP/InteropAPI.h
similarity index 100%
rename from openmp/libomptarget/include/OpenMP/InteropAPI.h
rename to offload/include/OpenMP/InteropAPI.h
diff --git a/openmp/libomptarget/include/OpenMP/Mapping.h b/offload/include/OpenMP/Mapping.h
similarity index 100%
rename from openmp/libomptarget/include/OpenMP/Mapping.h
rename to offload/include/OpenMP/Mapping.h
diff --git a/openmp/libomptarget/include/OpenMP/OMPT/Callback.h b/offload/include/OpenMP/OMPT/Callback.h
similarity index 100%
rename from openmp/libomptarget/include/OpenMP/OMPT/Callback.h
rename to offload/include/OpenMP/OMPT/Callback.h
diff --git a/openmp/libomptarget/include/OpenMP/OMPT/Connector.h b/offload/include/OpenMP/OMPT/Connector.h
similarity index 100%
rename from openmp/libomptarget/include/OpenMP/OMPT/Connector.h
rename to offload/include/OpenMP/OMPT/Connector.h
diff --git a/openmp/libomptarget/include/OpenMP/OMPT/Interface.h b/offload/include/OpenMP/OMPT/Interface.h
similarity index 100%
rename from openmp/libomptarget/include/OpenMP/OMPT/Interface.h
rename to offload/include/OpenMP/OMPT/Interface.h
diff --git a/openmp/libomptarget/include/OpenMP/omp.h b/offload/include/OpenMP/omp.h
similarity index 100%
rename from openmp/libomptarget/include/OpenMP/omp.h
rename to offload/include/OpenMP/omp.h
diff --git a/openmp/libomptarget/include/PluginManager.h b/offload/include/PluginManager.h
similarity index 100%
rename from openmp/libomptarget/include/PluginManager.h
rename to offload/include/PluginManager.h
diff --git a/openmp/libomptarget/include/Shared/APITypes.h b/offload/include/Shared/APITypes.h
similarity index 100%
rename from openmp/libomptarget/include/Shared/APITypes.h
rename to offload/include/Shared/APITypes.h
diff --git a/openmp/libomptarget/include/Shared/Debug.h b/offload/include/Shared/Debug.h
similarity index 100%
rename from openmp/libomptarget/include/Shared/Debug.h
rename to offload/include/Shared/Debug.h
diff --git a/openmp/libomptarget/include/Shared/Environment.h b/offload/include/Shared/Environment.h
similarity index 100%
rename from openmp/libomptarget/include/Shared/Environment.h
rename to offload/include/Shared/Environment.h
diff --git a/openmp/libomptarget/include/Shared/EnvironmentVar.h b/offload/include/Shared/EnvironmentVar.h
similarity index 100%
rename from openmp/libomptarget/include/Shared/EnvironmentVar.h
rename to offload/include/Shared/EnvironmentVar.h
diff --git a/openmp/libomptarget/include/Shared/PluginAPI.h b/offload/include/Shared/PluginAPI.h
similarity index 100%
rename from openmp/libomptarget/include/Shared/PluginAPI.h
rename to offload/include/Shared/PluginAPI.h
diff --git a/openmp/libomptarget/include/Shared/PluginAPI.inc b/offload/include/Shared/PluginAPI.inc
similarity index 100%
rename from openmp/libomptarget/include/Shared/PluginAPI.inc
rename to offload/include/Shared/PluginAPI.inc
diff --git a/openmp/libomptarget/include/Shared/Profile.h b/offload/include/Shared/Profile.h
similarity index 100%
rename from openmp/libomptarget/include/Shared/Profile.h
rename to offload/include/Shared/Profile.h
diff --git a/openmp/libomptarget/include/Shared/Requirements.h b/offload/include/Shared/Requirements.h
similarity index 100%
rename from openmp/libomptarget/include/Shared/Requirements.h
rename to offload/include/Shared/Requirements.h
diff --git a/openmp/libomptarget/include/Shared/SourceInfo.h b/offload/include/Shared/SourceInfo.h
similarity index 100%
rename from openmp/libomptarget/include/Shared/SourceInfo.h
rename to offload/include/Shared/SourceInfo.h
diff --git a/openmp/libomptarget/include/Shared/Utils.h b/offload/include/Shared/Utils.h
similarity index 100%
rename from openmp/libomptarget/include/Shared/Utils.h
rename to offload/include/Shared/Utils.h
diff --git a/openmp/libomptarget/include/Utils/ExponentialBackoff.h b/offload/include/Utils/ExponentialBackoff.h
similarity index 100%
rename from openmp/libomptarget/include/Utils/ExponentialBackoff.h
rename to offload/include/Utils/ExponentialBackoff.h
diff --git a/openmp/libomptarget/include/device.h b/offload/include/device.h
similarity index 100%
rename from openmp/libomptarget/include/device.h
rename to offload/include/device.h
diff --git a/openmp/libomptarget/include/omptarget.h b/offload/include/omptarget.h
similarity index 100%
rename from openmp/libomptarget/include/omptarget.h
rename to offload...
[truncated]

Copy link

github-actions bot commented Dec 12, 2023

⚠️ C/C++ code formatter, clang-format found issues in your code. ⚠️

You can test this locally with the following command:
git-clang-format --diff 446371f5ccbd81aa7707b4f1a1baa0ea0955dc67 c6a443db06e19ee114f74d847f52b1411df95833 -- offload/DeviceRTL/include/Allocator.h offload/DeviceRTL/include/Configuration.h offload/DeviceRTL/include/Debug.h offload/DeviceRTL/include/Interface.h offload/DeviceRTL/include/LibC.h offload/DeviceRTL/include/Mapping.h offload/DeviceRTL/include/State.h offload/DeviceRTL/include/Synchronization.h offload/DeviceRTL/include/Types.h offload/DeviceRTL/include/Utils.h offload/DeviceRTL/src/Allocator.cpp offload/DeviceRTL/src/Configuration.cpp offload/DeviceRTL/src/Debug.cpp offload/DeviceRTL/src/Kernel.cpp offload/DeviceRTL/src/LibC.cpp offload/DeviceRTL/src/Mapping.cpp offload/DeviceRTL/src/Misc.cpp offload/DeviceRTL/src/Parallelism.cpp offload/DeviceRTL/src/Reduction.cpp offload/DeviceRTL/src/State.cpp offload/DeviceRTL/src/Stub.cpp offload/DeviceRTL/src/Synchronization.cpp offload/DeviceRTL/src/Tasking.cpp offload/DeviceRTL/src/Utils.cpp offload/DeviceRTL/src/Workshare.cpp offload/include/DeviceImage.h offload/include/ExclusiveAccess.h offload/include/OffloadEntry.h offload/include/OffloadPolicy.h offload/include/OpenMP/InternalTypes.h offload/include/OpenMP/InteropAPI.h offload/include/OpenMP/Mapping.h offload/include/OpenMP/OMPT/Callback.h offload/include/OpenMP/OMPT/Connector.h offload/include/OpenMP/OMPT/Interface.h offload/include/OpenMP/omp.h offload/include/PluginManager.h offload/include/Shared/APITypes.h offload/include/Shared/Debug.h offload/include/Shared/Environment.h offload/include/Shared/EnvironmentVar.h offload/include/Shared/PluginAPI.h offload/include/Shared/PluginAPI.inc offload/include/Shared/Profile.h offload/include/Shared/Requirements.h offload/include/Shared/SourceInfo.h offload/include/Shared/Utils.h offload/include/Utils/ExponentialBackoff.h offload/include/device.h offload/include/omptarget.h offload/include/rtl.h offload/plugins-nextgen/amdgpu/dynamic_hsa/hsa.cpp offload/plugins-nextgen/amdgpu/dynamic_hsa/hsa.h offload/plugins-nextgen/amdgpu/dynamic_hsa/hsa_ext_amd.h offload/plugins-nextgen/amdgpu/src/rtl.cpp offload/plugins-nextgen/amdgpu/utils/UtilitiesRTL.h offload/plugins-nextgen/common/OMPT/OmptCallback.cpp offload/plugins-nextgen/common/include/DLWrap.h offload/plugins-nextgen/common/include/GlobalHandler.h offload/plugins-nextgen/common/include/JIT.h offload/plugins-nextgen/common/include/MemoryManager.h offload/plugins-nextgen/common/include/PluginInterface.h offload/plugins-nextgen/common/include/RPC.h offload/plugins-nextgen/common/include/Utils/ELF.h offload/plugins-nextgen/common/src/GlobalHandler.cpp offload/plugins-nextgen/common/src/JIT.cpp offload/plugins-nextgen/common/src/PluginInterface.cpp offload/plugins-nextgen/common/src/RPC.cpp offload/plugins-nextgen/common/src/Utils/ELF.cpp offload/plugins-nextgen/cuda/dynamic_cuda/cuda.cpp offload/plugins-nextgen/cuda/dynamic_cuda/cuda.h offload/plugins-nextgen/cuda/src/rtl.cpp offload/plugins-nextgen/host/dynamic_ffi/ffi.cpp offload/plugins-nextgen/host/dynamic_ffi/ffi.h offload/plugins-nextgen/host/src/rtl.cpp offload/src/DeviceImage.cpp offload/src/LegacyAPI.cpp offload/src/OffloadRTL.cpp offload/src/OpenMP/API.cpp offload/src/OpenMP/InteropAPI.cpp offload/src/OpenMP/Mapping.cpp offload/src/OpenMP/OMPT/Callback.cpp offload/src/PluginManager.cpp offload/src/device.cpp offload/src/interface.cpp offload/src/omptarget.cpp offload/src/private.h offload/test/Inputs/declare_indirect_func.c offload/test/api/assert.c offload/test/api/is_initial_device.c offload/test/api/omp_device_managed_memory.c offload/test/api/omp_device_managed_memory_alloc.c offload/test/api/omp_device_memory.c offload/test/api/omp_dynamic_shared_memory.c offload/test/api/omp_dynamic_shared_memory_amdgpu.c offload/test/api/omp_dynamic_shared_memory_mixed.inc offload/test/api/omp_dynamic_shared_memory_mixed_amdgpu.c offload/test/api/omp_dynamic_shared_memory_mixed_nvptx.c offload/test/api/omp_env_vars.c offload/test/api/omp_get_device_num.c offload/test/api/omp_get_mapped_ptr.c offload/test/api/omp_get_num_devices.c offload/test/api/omp_get_num_devices_with_empty_target.c offload/test/api/omp_get_num_procs.c offload/test/api/omp_host_pinned_memory.c offload/test/api/omp_host_pinned_memory_alloc.c offload/test/api/omp_indirect_call.c offload/test/api/omp_target_memcpy_async1.c offload/test/api/omp_target_memcpy_async2.c offload/test/api/omp_target_memcpy_rect_async1.c offload/test/api/omp_target_memcpy_rect_async2.c offload/test/api/omp_target_memset.c offload/test/api/ompx_3d.c offload/test/api/ompx_3d.cpp offload/test/api/ompx_sync.c offload/test/api/ompx_sync.cpp offload/test/env/base_ptr_ref_count.c offload/test/env/omp_target_debug.c offload/test/jit/empty_kernel.inc offload/test/jit/empty_kernel_lvl1.c offload/test/jit/empty_kernel_lvl2.c offload/test/jit/type_punning.c offload/test/libc/assert.c offload/test/libc/fwrite.c offload/test/libc/global_ctor_dtor.cpp offload/test/libc/host_call.c offload/test/libc/malloc.c offload/test/libc/puts.c offload/test/mapping/alloc_fail.c offload/test/mapping/array_section_implicit_capture.c offload/test/mapping/array_section_use_device_ptr.c offload/test/mapping/auto_zero_copy.cpp offload/test/mapping/auto_zero_copy_apu.cpp offload/test/mapping/auto_zero_copy_globals.cpp offload/test/mapping/data_absent_at_exit.c offload/test/mapping/data_member_ref.cpp offload/test/mapping/declare_mapper_api.cpp offload/test/mapping/declare_mapper_nested_default_mappers.cpp offload/test/mapping/declare_mapper_nested_default_mappers_array.cpp offload/test/mapping/declare_mapper_nested_default_mappers_array_subscript.cpp offload/test/mapping/declare_mapper_nested_default_mappers_complex_structure.cpp offload/test/mapping/declare_mapper_nested_default_mappers_ptr_subscript.cpp offload/test/mapping/declare_mapper_nested_default_mappers_var.cpp offload/test/mapping/declare_mapper_nested_mappers.cpp offload/test/mapping/declare_mapper_target.cpp offload/test/mapping/declare_mapper_target_data.cpp offload/test/mapping/declare_mapper_target_data_enter_exit.cpp offload/test/mapping/declare_mapper_target_update.cpp offload/test/mapping/delete_inf_refcount.c offload/test/mapping/device_ptr_update.c offload/test/mapping/firstprivate_aligned.cpp offload/test/mapping/has_device_addr.cpp offload/test/mapping/implicit_device_ptr.c offload/test/mapping/is_device_ptr.cpp offload/test/mapping/lambda_by_value.cpp offload/test/mapping/lambda_mapping.cpp offload/test/mapping/low_alignment.c offload/test/mapping/map_back_race.cpp offload/test/mapping/ompx_hold/omp_target_disassociate_ptr.c offload/test/mapping/ompx_hold/struct.c offload/test/mapping/ompx_hold/target-data.c offload/test/mapping/ompx_hold/target.c offload/test/mapping/padding_not_mapped.c offload/test/mapping/power_of_two_alignment.c offload/test/mapping/pr38704.c offload/test/mapping/prelock.cpp offload/test/mapping/present/target.c offload/test/mapping/present/target_array_extension.c offload/test/mapping/present/target_data.c offload/test/mapping/present/target_data_array_extension.c offload/test/mapping/present/target_data_at_exit.c offload/test/mapping/present/target_enter_data.c offload/test/mapping/present/target_exit_data_delete.c offload/test/mapping/present/target_exit_data_release.c offload/test/mapping/present/target_update.c offload/test/mapping/present/target_update_array_extension.c offload/test/mapping/present/unified_shared_memory.c offload/test/mapping/present/zero_length_array_section.c offload/test/mapping/present/zero_length_array_section_exit.c offload/test/mapping/private_mapping.c offload/test/mapping/ptr_and_obj_motion.c offload/test/mapping/reduction_implicit_map.cpp offload/test/mapping/target_data_array_extension_at_exit.c offload/test/mapping/target_derefence_array_pointrs.cpp offload/test/mapping/target_has_device_addr.c offload/test/mapping/target_implicit_partial_map.c offload/test/mapping/target_map_for_member_data.cpp offload/test/mapping/target_pointers_members_map.cpp offload/test/mapping/target_update_array_extension.c offload/test/mapping/target_use_device_addr.c offload/test/mapping/target_uses_allocator.c offload/test/mapping/target_wrong_use_device_addr.c offload/test/offloading/assert.cpp offload/test/offloading/atomic-compare-signedness.c offload/test/offloading/back2back_distribute.c offload/test/offloading/barrier_fence.c offload/test/offloading/bug47654.cpp offload/test/offloading/bug49021.cpp offload/test/offloading/bug49334.cpp offload/test/offloading/bug49779.cpp offload/test/offloading/bug50022.cpp offload/test/offloading/bug51781.c offload/test/offloading/bug51982.c offload/test/offloading/bug53727.cpp offload/test/offloading/bug64959.c offload/test/offloading/bug64959_compile_only.c offload/test/offloading/bug74582.c offload/test/offloading/complex_reduction.cpp offload/test/offloading/ctor_dtor.cpp offload/test/offloading/cuda_no_devices.c offload/test/offloading/d2d_memcpy.c offload/test/offloading/d2d_memcpy_sync.c offload/test/offloading/default_thread_limit.c offload/test/offloading/dynamic_module.c offload/test/offloading/dynamic_module_load.c offload/test/offloading/extern.c offload/test/offloading/force-usm.cpp offload/test/offloading/fortran/basic_array.c offload/test/offloading/generic_multiple_parallel_regions.c offload/test/offloading/global_constructor.cpp offload/test/offloading/host_as_target.c offload/test/offloading/indirect_fp_mapping.c offload/test/offloading/info.c offload/test/offloading/interop.c offload/test/offloading/lone_target_exit_data.c offload/test/offloading/looptripcnt.c offload/test/offloading/malloc.c offload/test/offloading/malloc_parallel.c offload/test/offloading/mandatory_but_no_devices.c offload/test/offloading/memory_manager.cpp offload/test/offloading/multiple_reductions_simple.c offload/test/offloading/non_contiguous_update.cpp offload/test/offloading/offloading_success.c offload/test/offloading/offloading_success.cpp offload/test/offloading/ompx_bare.c offload/test/offloading/ompx_coords.c offload/test/offloading/ompx_saxpy_mixed.c offload/test/offloading/parallel_offloading_map.cpp offload/test/offloading/parallel_target_teams_reduction.cpp offload/test/offloading/parallel_target_teams_reduction_max.cpp offload/test/offloading/parallel_target_teams_reduction_min.cpp offload/test/offloading/requires.c offload/test/offloading/runtime_init.c offload/test/offloading/shared_lib_fp_mapping.c offload/test/offloading/small_trip_count.c offload/test/offloading/small_trip_count_thread_limit.cpp offload/test/offloading/spmdization.c offload/test/offloading/static_linking.c offload/test/offloading/std_complex_arithmetic.cpp offload/test/offloading/struct_mapping_with_pointers.cpp offload/test/offloading/target-teams-atomic.c offload/test/offloading/target-tile.c offload/test/offloading/target_constexpr_mapping.cpp offload/test/offloading/target_critical_region.cpp offload/test/offloading/target_depend_nowait.cpp offload/test/offloading/target_map_for_member_data.cpp offload/test/offloading/target_nowait_target.cpp offload/test/offloading/task_in_reduction_target.c offload/test/offloading/taskloop_offload_nowait.cpp offload/test/offloading/test_libc.cpp offload/test/offloading/thread_limit.c offload/test/offloading/thread_state_1.c offload/test/offloading/thread_state_2.c offload/test/offloading/weak.c offload/test/offloading/workshare_chunk.c offload/test/offloading/wtime.c offload/test/ompt/callbacks.h offload/test/ompt/register_both.h offload/test/ompt/register_emi.h offload/test/ompt/register_emi_map.h offload/test/ompt/register_no_device_init.h offload/test/ompt/register_non_emi.h offload/test/ompt/register_non_emi_map.h offload/test/ompt/register_wrong_return.h offload/test/ompt/target_memcpy.c offload/test/ompt/target_memcpy_emi.c offload/test/ompt/veccopy.c offload/test/ompt/veccopy_data.c offload/test/ompt/veccopy_disallow_both.c offload/test/ompt/veccopy_emi.c offload/test/ompt/veccopy_emi_map.c offload/test/ompt/veccopy_map.c offload/test/ompt/veccopy_no_device_init.c offload/test/ompt/veccopy_wrong_return.c offload/test/unified_shared_memory/api.c offload/test/unified_shared_memory/associate_ptr.c offload/test/unified_shared_memory/close_enter_exit.c offload/test/unified_shared_memory/close_manual.c offload/test/unified_shared_memory/close_member.c offload/test/unified_shared_memory/close_modifier.c offload/test/unified_shared_memory/shared_update.c offload/tools/deviceinfo/llvm-omp-device-info.cpp offload/tools/kernelreplay/llvm-omp-kernel-replay.cpp offload/unittests/Plugins/NextgenPluginsTest.cpp
View the diff from clang-format here.
diff --git a/offload/DeviceRTL/include/Types.h b/offload/DeviceRTL/include/Types.h
index 2e12d9da03..6d32f100c4 100644
--- a/offload/DeviceRTL/include/Types.h
+++ b/offload/DeviceRTL/include/Types.h
@@ -115,9 +115,9 @@ enum kmp_sched_t {
 #define SCHEDULE_WITHOUT_MODIFIERS(s)                                          \
   (enum kmp_sched_t)(                                                          \
       (s) & ~(kmp_sched_modifier_nonmonotonic | kmp_sched_modifier_monotonic))
-#define SCHEDULE_HAS_MONOTONIC(s) (((s)&kmp_sched_modifier_monotonic) != 0)
+#define SCHEDULE_HAS_MONOTONIC(s) (((s) & kmp_sched_modifier_monotonic) != 0)
 #define SCHEDULE_HAS_NONMONOTONIC(s)                                           \
-  (((s)&kmp_sched_modifier_nonmonotonic) != 0)
+  (((s) & kmp_sched_modifier_nonmonotonic) != 0)
 #define SCHEDULE_HAS_NO_MODIFIERS(s)                                           \
   (((s) & (kmp_sched_modifier_nonmonotonic | kmp_sched_modifier_monotonic)) == \
    0)
diff --git a/offload/include/OpenMP/OMPT/Interface.h b/offload/include/OpenMP/OMPT/Interface.h
index 327fadfcd4..4f624bd00b 100644
--- a/offload/include/OpenMP/OMPT/Interface.h
+++ b/offload/include/OpenMP/OMPT/Interface.h
@@ -254,8 +254,8 @@ private:
 
 // InterfaceRAII's class template argument deduction guide
 template <typename CallbackPairTy, typename... ArgsTy>
-InterfaceRAII(CallbackPairTy Callbacks, ArgsTy... Args)
-    -> InterfaceRAII<CallbackPairTy, ArgsTy...>;
+InterfaceRAII(CallbackPairTy Callbacks,
+              ArgsTy... Args) -> InterfaceRAII<CallbackPairTy, ArgsTy...>;
 
 /// Used to set and reset the thread-local return address. The RAII is expected
 /// to be created at a runtime entry point when the return address should be
diff --git a/offload/include/OpenMP/omp.h b/offload/include/OpenMP/omp.h
index b44c6aff1b..201ec5acaf 100644
--- a/offload/include/OpenMP/omp.h
+++ b/offload/include/OpenMP/omp.h
@@ -101,8 +101,9 @@ int __KAI_KMPC_CONVENTION omp_get_num_interop_properties(const omp_interop_t);
  * The `omp_get_interop_int` routine retrieves an integer property from an
  * `omp_interop_t` object.
  */
-omp_intptr_t __KAI_KMPC_CONVENTION
-omp_get_interop_int(const omp_interop_t, omp_interop_property_t, int *);
+omp_intptr_t __KAI_KMPC_CONVENTION omp_get_interop_int(const omp_interop_t,
+                                                       omp_interop_property_t,
+                                                       int *);
 /*!
  * The `omp_get_interop_ptr` routine retrieves a pointer property from an
  * `omp_interop_t` object.
@@ -113,8 +114,9 @@ void *__KAI_KMPC_CONVENTION omp_get_interop_ptr(const omp_interop_t,
  * The `omp_get_interop_str` routine retrieves a string property from an
  * `omp_interop_t` object.
  */
-const char *__KAI_KMPC_CONVENTION
-omp_get_interop_str(const omp_interop_t, omp_interop_property_t, int *);
+const char *__KAI_KMPC_CONVENTION omp_get_interop_str(const omp_interop_t,
+                                                      omp_interop_property_t,
+                                                      int *);
 /*!
  * The `omp_get_interop_name` routine retrieves a property name from an
  * `omp_interop_t` object.
diff --git a/offload/include/Shared/APITypes.h b/offload/include/Shared/APITypes.h
index e8fc27785b..2ebd03af8e 100644
--- a/offload/include/Shared/APITypes.h
+++ b/offload/include/Shared/APITypes.h
@@ -26,11 +26,11 @@ extern "C" {
 /// This struct is a record of an entry point or global. For a function
 /// entry point the size is expected to be zero
 struct __tgt_offload_entry {
-  void *addr;       // Pointer to the offload entry info (function or global)
-  char *name;       // Name of the function or global
-  size_t size;      // Size of the entry info (0 if it is a function)
-  int32_t flags;    // Flags associated with the entry, e.g. 'link'.
-  int32_t data;     // Extra data associated with the entry.
+  void *addr;    // Pointer to the offload entry info (function or global)
+  char *name;    // Name of the function or global
+  size_t size;   // Size of the entry info (0 if it is a function)
+  int32_t flags; // Flags associated with the entry, e.g. 'link'.
+  int32_t data;  // Extra data associated with the entry.
 };
 
 /// This struct is a record of the device image information
diff --git a/offload/plugins-nextgen/amdgpu/dynamic_hsa/hsa_ext_amd.h b/offload/plugins-nextgen/amdgpu/dynamic_hsa/hsa_ext_amd.h
index 3117763e35..82c607f474 100644
--- a/offload/plugins-nextgen/amdgpu/dynamic_hsa/hsa_ext_amd.h
+++ b/offload/plugins-nextgen/amdgpu/dynamic_hsa/hsa_ext_amd.h
@@ -104,11 +104,11 @@ hsa_status_t hsa_amd_agents_allow_access(uint32_t num_agents,
                                          const uint32_t *flags,
                                          const void *ptr);
 
-hsa_status_t hsa_amd_memory_lock(void* host_ptr, size_t size,
-                                hsa_agent_t* agents, int num_agent,
-                                void** agent_ptr);
+hsa_status_t hsa_amd_memory_lock(void *host_ptr, size_t size,
+                                 hsa_agent_t *agents, int num_agent,
+                                 void **agent_ptr);
 
-hsa_status_t hsa_amd_memory_unlock(void* host_ptr);
+hsa_status_t hsa_amd_memory_unlock(void *host_ptr);
 
 hsa_status_t hsa_amd_memory_fill(void *ptr, uint32_t value, size_t count);
 
@@ -156,16 +156,15 @@ typedef enum {
 typedef struct hsa_amd_pointer_info_s {
   uint32_t size;
   hsa_amd_pointer_type_t type;
-  void* agentBaseAddress;
-  void* hostBaseAddress;
+  void *agentBaseAddress;
+  void *hostBaseAddress;
   size_t sizeInBytes;
 } hsa_amd_pointer_info_t;
 
-hsa_status_t hsa_amd_pointer_info(const void* ptr,
-                                          hsa_amd_pointer_info_t* info,
-                                          void* (*alloc)(size_t),
-                                          uint32_t* num_agents_accessible,
-                                          hsa_agent_t** accessible);
+hsa_status_t hsa_amd_pointer_info(const void *ptr, hsa_amd_pointer_info_t *info,
+                                  void *(*alloc)(size_t),
+                                  uint32_t *num_agents_accessible,
+                                  hsa_agent_t **accessible);
 
 #ifdef __cplusplus
 }
diff --git a/offload/plugins-nextgen/common/include/DLWrap.h b/offload/plugins-nextgen/common/include/DLWrap.h
index 8934e7e701..00610678f9 100644
--- a/offload/plugins-nextgen/common/include/DLWrap.h
+++ b/offload/plugins-nextgen/common/include/DLWrap.h
@@ -107,7 +107,9 @@ template <size_t S> struct count {
   static constexpr size_t N = count<S - 1>::N;
 };
 
-template <> struct count<0> { static constexpr size_t N = 0; };
+template <> struct count<0> {
+  static constexpr size_t N = 0;
+};
 
 // Get a constexpr size_t ID, starts at zero
 #define DLWRAP_ID() (dlwrap::type::count<__LINE__>::N)
diff --git a/offload/plugins-nextgen/cuda/dynamic_cuda/cuda.h b/offload/plugins-nextgen/cuda/dynamic_cuda/cuda.h
index 32031c28f8..96c22dda1e 100644
--- a/offload/plugins-nextgen/cuda/dynamic_cuda/cuda.h
+++ b/offload/plugins-nextgen/cuda/dynamic_cuda/cuda.h
@@ -24,12 +24,12 @@ typedef struct CUfunc_st *CUfunction;
 typedef struct CUstream_st *CUstream;
 typedef struct CUevent_st *CUevent;
 
-#define CU_DEVICE_INVALID ((CUdevice)-2)
+#define CU_DEVICE_INVALID ((CUdevice) - 2)
 
 typedef unsigned long long CUmemGenericAllocationHandle_v1;
 typedef CUmemGenericAllocationHandle_v1 CUmemGenericAllocationHandle;
 
-#define CU_DEVICE_INVALID ((CUdevice)-2)
+#define CU_DEVICE_INVALID ((CUdevice) - 2)
 
 typedef enum CUmemAllocationGranularity_flags_enum {
   CU_MEM_ALLOC_GRANULARITY_MINIMUM = 0x0,
diff --git a/offload/src/OpenMP/API.cpp b/offload/src/OpenMP/API.cpp
index c85f9868e3..77bdabb52f 100644
--- a/offload/src/OpenMP/API.cpp
+++ b/offload/src/OpenMP/API.cpp
@@ -644,8 +644,8 @@ EXTERN void *omp_get_mapped_ptr(const void *Ptr, int DeviceNum) {
 
   size_t NumDevices = omp_get_initial_device();
   if (DeviceNum == NumDevices) {
-    DP("Device %d is initial device, returning Ptr " DPxMOD ".\n",
-           DeviceNum, DPxPTR(Ptr));
+    DP("Device %d is initial device, returning Ptr " DPxMOD ".\n", DeviceNum,
+       DPxPTR(Ptr));
     return const_cast<void *>(Ptr);
   }
 
diff --git a/offload/test/api/omp_dynamic_shared_memory_amdgpu.c b/offload/test/api/omp_dynamic_shared_memory_amdgpu.c
index 0b4d9d6ea9..c4d7c8da2e 100644
--- a/offload/test/api/omp_dynamic_shared_memory_amdgpu.c
+++ b/offload/test/api/omp_dynamic_shared_memory_amdgpu.c
@@ -1,4 +1,5 @@
-// RUN: %libomptarget-compile-amdgcn-amd-amdhsa -O1 -mllvm -openmp-opt-inline-device
+// RUN: %libomptarget-compile-amdgcn-amd-amdhsa -O1 -mllvm
+// -openmp-opt-inline-device
 // RUN: env LIBOMPTARGET_SHARED_MEMORY_SIZE=256 \
 // RUN:   %libomptarget-run-amdgcn-amd-amdhsa | %fcheck-amdgcn-amd-amdhsa
 // REQUIRES: amdgcn-amd-amdhsa
diff --git a/offload/test/api/omp_dynamic_shared_memory_mixed.inc b/offload/test/api/omp_dynamic_shared_memory_mixed.inc
index 16e0becd09..08e050b935 100644
--- a/offload/test/api/omp_dynamic_shared_memory_mixed.inc
+++ b/offload/test/api/omp_dynamic_shared_memory_mixed.inc
@@ -9,8 +9,7 @@ int main() {
   int Result[N], NumThreads;
 
 #pragma omp target teams num_teams(1) thread_limit(N)                          \
-                         ompx_dyn_cgroup_mem(N * sizeof(Result[0]))            \
-                         map(from : Result, NumThreads)
+    ompx_dyn_cgroup_mem(N * sizeof(Result[0])) map(from : Result, NumThreads)
   {
     int Buffer[N];
 #pragma omp parallel
diff --git a/offload/test/api/omp_dynamic_shared_memory_mixed_amdgpu.c b/offload/test/api/omp_dynamic_shared_memory_mixed_amdgpu.c
index 656c3a20aa..9de28974d6 100644
--- a/offload/test/api/omp_dynamic_shared_memory_mixed_amdgpu.c
+++ b/offload/test/api/omp_dynamic_shared_memory_mixed_amdgpu.c
@@ -1,4 +1,5 @@
-// RUN: %libomptarget-compile-amdgcn-amd-amdhsa -O1 -mllvm -openmp-opt-inline-device -I %S
+// RUN: %libomptarget-compile-amdgcn-amd-amdhsa -O1 -mllvm
+// -openmp-opt-inline-device -I %S
 // RUN: env LIBOMPTARGET_NEXTGEN_PLUGINS=1 \
 // RUN:   %libomptarget-run-amdgcn-amd-amdhsa | %fcheck-amdgcn-amd-amdhsa
 // REQUIRES: amdgcn-amd-amdhsa
diff --git a/offload/test/api/omp_get_mapped_ptr.c b/offload/test/api/omp_get_mapped_ptr.c
index a8e11f912d..6d6fe42bc3 100644
--- a/offload/test/api/omp_get_mapped_ptr.c
+++ b/offload/test/api/omp_get_mapped_ptr.c
@@ -13,7 +13,7 @@ int main(int argc, char *argv[]) {
 
   assert(device_ptr == NULL && "the pointer should not be mapped right now");
 
-#pragma omp target enter data map(to: host_data[:N])
+#pragma omp target enter data map(to : host_data[ : N])
 
   device_ptr = omp_get_mapped_ptr(host_data, 0);
 
@@ -21,7 +21,7 @@ int main(int argc, char *argv[]) {
 
   void *ptr = NULL;
 
-#pragma omp target map(from: ptr)
+#pragma omp target map(from : ptr)
   { ptr = host_data; }
 
   assert(ptr == device_ptr && "wrong pointer mapping");
@@ -30,7 +30,7 @@ int main(int argc, char *argv[]) {
 
   assert(device_ptr && "the pointer with offset should be mapped");
 
-#pragma omp target map(from: ptr)
+#pragma omp target map(from : ptr)
   { ptr = host_data + OFFSET; }
 
   assert(ptr == device_ptr && "wrong pointer mapping");
diff --git a/offload/test/api/ompx_3d.c b/offload/test/api/ompx_3d.c
index 2ae17c226f..3045c1d2fc 100644
--- a/offload/test/api/ompx_3d.c
+++ b/offload/test/api/ompx_3d.c
@@ -6,8 +6,7 @@
 
 void foo(int device) {
   int tid = 0, bid = 0, bdim = 0;
-#pragma omp target teams distribute parallel for map(from                      \
-                                                     : tid, bid, bdim)         \
+#pragma omp target teams distribute parallel for map(from : tid, bid, bdim)    \
     device(device) thread_limit(2) num_teams(5)
   for (int i = 0; i < 1000; ++i) {
     if (i == 42) {
diff --git a/offload/test/api/ompx_3d.cpp b/offload/test/api/ompx_3d.cpp
index 5b5491263a..f25b851532 100644
--- a/offload/test/api/ompx_3d.cpp
+++ b/offload/test/api/ompx_3d.cpp
@@ -6,8 +6,7 @@
 
 void foo(int device) {
   int tid = 0, bid = 0, bdim = 0;
-#pragma omp target teams distribute parallel for map(from                      \
-                                                     : tid, bid, bdim)         \
+#pragma omp target teams distribute parallel for map(from : tid, bid, bdim)    \
     device(device) thread_limit(2) num_teams(5)
   for (int i = 0; i < 1000; ++i) {
     if (i == 42) {
diff --git a/offload/test/env/base_ptr_ref_count.c b/offload/test/env/base_ptr_ref_count.c
index 9a15568712..876991176d 100644
--- a/offload/test/env/base_ptr_ref_count.c
+++ b/offload/test/env/base_ptr_ref_count.c
@@ -1,5 +1,5 @@
-// RUN: %libomptarget-compile-generic && env LIBOMPTARGET_DEBUG=1 %libomptarget-run-generic 2>&1 | %fcheck-generic
-// REQUIRES: libomptarget-debug
+// RUN: %libomptarget-compile-generic && env LIBOMPTARGET_DEBUG=1
+// %libomptarget-run-generic 2>&1 | %fcheck-generic REQUIRES: libomptarget-debug
 
 #include <stdio.h>
 #include <stdlib.h>
diff --git a/offload/test/env/omp_target_debug.c b/offload/test/env/omp_target_debug.c
index ec81873a09..8694448acb 100644
--- a/offload/test/env/omp_target_debug.c
+++ b/offload/test/env/omp_target_debug.c
@@ -1,6 +1,8 @@
-// RUN: %libomptarget-compile-generic && env LIBOMPTARGET_DEBUG=1 %libomptarget-run-generic 2>&1 | %fcheck-generic -allow-empty -check-prefix=DEBUG
-// RUN: %libomptarget-compile-generic && env LIBOMPTARGET_DEBUG=0 %libomptarget-run-generic 2>&1 | %fcheck-generic -allow-empty -check-prefix=NDEBUG
-// REQUIRES: libomptarget-debug
+// RUN: %libomptarget-compile-generic && env LIBOMPTARGET_DEBUG=1
+// %libomptarget-run-generic 2>&1 | %fcheck-generic -allow-empty
+// -check-prefix=DEBUG RUN: %libomptarget-compile-generic && env
+// LIBOMPTARGET_DEBUG=0 %libomptarget-run-generic 2>&1 | %fcheck-generic
+// -allow-empty -check-prefix=NDEBUG REQUIRES: libomptarget-debug
 
 int main(void) {
 #pragma omp target
diff --git a/offload/test/jit/empty_kernel.inc b/offload/test/jit/empty_kernel.inc
index 43813891cc..3e60ca994a 100644
--- a/offload/test/jit/empty_kernel.inc
+++ b/offload/test/jit/empty_kernel.inc
@@ -1,15 +1,14 @@
-int main(int argc, char** argv) {
-  #pragma omp TGT1_DIRECTIVE
+int main(int argc, char **argv) {
+#pragma omp TGT1_DIRECTIVE
   {
 #ifdef LOOP_DIRECTIVE
-    #pragma omp LOOP_DIRECTIVE
+#pragma omp LOOP_DIRECTIVE
     for (int i = 0; i < argc; ++i)
 #endif
     {
 #ifdef BODY_DIRECTIVE
-      #pragma omp BODY_DIRECTIVE
-      {
-      }
+#pragma omp BODY_DIRECTIVE
+      {}
 #endif
     }
   }
@@ -18,14 +17,13 @@ int main(int argc, char** argv) {
 #pragma omp TGT2_DIRECTIVE
   {
 #ifdef LOOP_DIRECTIVE
-    #pragma omp LOOP_DIRECTIVE
+#pragma omp LOOP_DIRECTIVE
     for (int i = 0; i < argc; ++i)
 #endif
     {
 #ifdef BODY_DIRECTIVE
-      #pragma omp BODY_DIRECTIVE
-      {
-      }
+#pragma omp BODY_DIRECTIVE
+      {}
 #endif
     }
   }
diff --git a/offload/test/mapping/alloc_fail.c b/offload/test/mapping/alloc_fail.c
index c4ae70fc73..31912d4bdb 100644
--- a/offload/test/mapping/alloc_fail.c
+++ b/offload/test/mapping/alloc_fail.c
@@ -2,9 +2,11 @@
 // RUN: %libomptarget-run-fail-generic 2>&1 \
 // RUN: | %fcheck-generic
 
-// CHECK: omptarget message: explicit extension not allowed: host address specified is 0x{{.*}} (8 bytes), but device allocation maps to host at 0x{{.*}} (8 bytes)
-// CHECK: omptarget error: Call to getTargetPointer returned null pointer (device failure or illegal mapping).
-// CHECK: omptarget fatal error 1: failure of target construct while offloading is mandatory
+// CHECK: omptarget message: explicit extension not allowed: host address
+// specified is 0x{{.*}} (8 bytes), but device allocation maps to host at
+// 0x{{.*}} (8 bytes) CHECK: omptarget error: Call to getTargetPointer returned
+// null pointer (device failure or illegal mapping). CHECK: omptarget fatal
+// error 1: failure of target construct while offloading is mandatory
 
 int main() {
   int arr[4] = {0, 1, 2, 3};
diff --git a/offload/test/mapping/array_section_implicit_capture.c b/offload/test/mapping/array_section_implicit_capture.c
index 210b7e51cb..1042bd23bd 100644
--- a/offload/test/mapping/array_section_implicit_capture.c
+++ b/offload/test/mapping/array_section_implicit_capture.c
@@ -1,4 +1,4 @@
-// RUN: %libomptarget-compile-generic 
+// RUN: %libomptarget-compile-generic
 // RUN: %libomptarget-run-generic 2>&1 \
 // RUN: | %fcheck-generic
 
diff --git a/offload/test/mapping/data_absent_at_exit.c b/offload/test/mapping/data_absent_at_exit.c
index 5baaf933b2..bbbff1e621 100644
--- a/offload/test/mapping/data_absent_at_exit.c
+++ b/offload/test/mapping/data_absent_at_exit.c
@@ -16,7 +16,8 @@ int main(void) {
 #pragma omp target data map(from : f) map(always, from : af)
   {
 #pragma omp target exit data map(delete : f, af)
-  } printf("f = %d, af = %d\n", f, af);
+  }
+  printf("f = %d, af = %d\n", f, af);
 
 // Check omp target exit data.
 // CHECK: f = 5, r = 6, d = 7, af = 8
diff --git a/offload/test/mapping/declare_mapper_nested_default_mappers.cpp b/offload/test/mapping/declare_mapper_nested_default_mappers.cpp
index c6c5657ae6..6408dae555 100644
--- a/offload/test/mapping/declare_mapper_nested_default_mappers.cpp
+++ b/offload/test/mapping/declare_mapper_nested_default_mappers.cpp
@@ -44,8 +44,8 @@ int main() {
 
   int spp00fa = -1, spp00fca = -1, spp00fb_r = -1;
   __intptr_t p = reinterpret_cast<__intptr_t>(&x[0]);
-#pragma omp target map(tofrom: spp[0][0]) firstprivate(p)                           \
-                   map(from: spp00fa, spp00fca, spp00fb_r)
+#pragma omp target map(tofrom : spp[0][0]) firstprivate(p)                     \
+    map(from : spp00fa, spp00fca, spp00fb_r)
   {
     spp00fa = spp[0][0].f.a;
     spp00fca = spp[0][0].f.c.a;
diff --git a/offload/test/mapping/declare_mapper_nested_mappers.cpp b/offload/test/mapping/declare_mapper_nested_mappers.cpp
index a9e3f05e0f..cbcb10bbf1 100644
--- a/offload/test/mapping/declare_mapper_nested_mappers.cpp
+++ b/offload/test/mapping/declare_mapper_nested_mappers.cpp
@@ -42,8 +42,8 @@ int main() {
   int spp00fa = -1, spp00fb_r = -1, spp00fg1 = -1, spp00fg_r = -1;
   __intptr_t p = reinterpret_cast<__intptr_t>(&x[0]),
              p1 = reinterpret_cast<__intptr_t>(&y[0]);
-#pragma omp target map(tofrom : spp[0][0]) firstprivate(p, p1)                  \
-                   map(from: spp00fa, spp00fb_r, spp00fg1, spp00fg_r)
+#pragma omp target map(tofrom : spp[0][0]) firstprivate(p, p1)                 \
+    map(from : spp00fa, spp00fb_r, spp00fg1, spp00fg_r)
   {
     spp00fa = spp[0][0].f.a;
     spp00fb_r = spp[0][0].f.b == reinterpret_cast<void *>(p) ? 1 : 0;
@@ -56,7 +56,7 @@ int main() {
     spp[0][0].f.b[1] = 40;
     spp[0][0].g[1] = 50;
   }
-    printf("%d %d %d %d\n", spp00fa, spp00fb_r, spp00fg1, spp00fg_r);
+  printf("%d %d %d %d\n", spp00fa, spp00fb_r, spp00fg1, spp00fg_r);
   // CHECK: 222 0 30 0
   printf("%d %d %4.5f %d %d %d\n", spp[0][0].e, spp[0][0].f.a, spp[0][0].f.b[1],
          spp[0][0].f.b == &x[0] ? 1 : 0, spp[0][0].g[1],
diff --git a/offload/test/mapping/device_ptr_update.c b/offload/test/mapping/device_ptr_update.c
index c6719d2285..77aab6c579 100644
--- a/offload/test/mapping/device_ptr_update.c
+++ b/offload/test/mapping/device_ptr_update.c
@@ -15,7 +15,8 @@ int main(void) {
 
   s1.p = A;
 
-// DEBUG: Update pointer ([[DEV_PTR:0x[^ ]+]]) -> {{\[}}[[DEV_OBJ_A:0x[^ ]+]]{{\]}}
+// DEBUG: Update pointer ([[DEV_PTR:0x[^ ]+]]) -> {{\[}}[[DEV_OBJ_A:0x[^
+// ]+]]{{\]}}
 #pragma omp target enter data map(alloc : s1.p[0 : 10])
 
 // DEBUG-NOT: Update pointer ([[DEV_PTR]]) -> {{\[}}[[DEV_OBJ_A]]{{\]}}
diff --git a/offload/test/mapping/padding_not_mapped.c b/offload/test/mapping/padding_not_mapped.c
index 3ee70ab640..99f57f4fbb 100644
--- a/offload/test/mapping/padding_not_mapped.c
+++ b/offload/test/mapping/padding_not_mapped.c
@@ -18,7 +18,11 @@
 #include <stdio.h>
 
 int main() {
-  struct S { int x; int y; double z; } s = {1, 2, 3};
+  struct S {
+    int x;
+    int y;
+    double z;
+  } s = {1, 2, 3};
 
   // CHECK: &s.x = 0x[[#%x,HOST_ADDR:]], size = [[#%u,SIZE:]]
   fprintf(stderr, "&s = %p\n", &s);
@@ -26,18 +30,20 @@ int main() {
   fprintf(stderr, "&s.y = %p\n", &s.y);
   fprintf(stderr, "&s.z = %p\n", &s.z);
 
-  // CHECK: s.x is present: 0
-  // CHECK: s.x = 1{{$}}
-  #pragma omp target enter data map(alloc: s.y, s.z)
+// CHECK: s.x is present: 0
+// CHECK: s.x = 1{{$}}
+#pragma omp target enter data map(alloc : s.y, s.z)
   int dev = omp_get_default_device();
   fprintf(stderr, "s.x is present: %d\n", omp_target_is_present(&s.x, dev));
-  #pragma omp target update from(s.x) // should have no effect
+#pragma omp target update from(s.x) // should have no effect
   fprintf(stderr, "s.x = %d\n", s.x);
 
-  // CHECK: omptarget message: device mapping required by 'present' map type modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]] bytes)
-  // CHECK: omptarget error: Call to getTargetPointer returned null pointer ('present' map type modifier).
-  // CHECK: omptarget fatal error 1: failure of target construct while offloading is mandatory
-  #pragma omp target enter data map(present, alloc: s.x)
+// CHECK: omptarget message: device mapping required by 'present' map type
+// modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]]
+// bytes) CHECK: omptarget error: Call to getTargetPointer returned null pointer
+// ('present' map type modifier). CHECK: omptarget fatal error 1: failure of
+// target construct while offloading is mandatory
+#pragma omp target enter data map(present, alloc : s.x)
 
   return 0;
 }
diff --git a/offload/test/mapping/power_of_two_alignment.c b/offload/test/mapping/power_of_two_alignment.c
index faebe4f89f..b31c43ac51 100644
--- a/offload/test/mapping/power_of_two_alignment.c
+++ b/offload/test/mapping/power_of_two_alignment.c
@@ -42,10 +42,13 @@
 #include <stdint.h>
 #include <stdio.h>
 
-template <typename StackPad>
-void test() {
+template <typename StackPad> void test() {
   StackPad stackPad;
-  struct S { char x; char y[7]; char z[8]; };
+  struct S {
+    char x;
+    char y[7];
+    char z[8];
+  };
   struct S collidePre, s, collidePost;
   uintptr_t mod16 = (uintptr_t)&s % 16;
   fprintf(stderr, "&s = %p\n", &s);
@@ -56,8 +59,8 @@ void test() {
   }
   fprintf(stderr, "&collidePre = %p\n", &collidePre);
   fprintf(stderr, "&collidePost = %p\n", &collidePost);
-  #pragma omp target data map(to:s.y, s.z)
-  #pragma omp target data map(to:collidePre, collidePost)
+#pragma omp target data map(to : s.y, s.z)
+#pragma omp target data map(to : collidePre, collidePost)
   ;
 }
 
diff --git a/offload/test/mapping/present/target.c b/offload/test/mapping/present/target.c
index 4344c42c80..e5cab61374 100644
--- a/offload/test/mapping/present/target.c
+++ b/offload/test/mapping/present/target.c
@@ -18,11 +18,13 @@ int main() {
   // CHECK: i is present
   fprintf(stderr, "i is present\n");
 
-  // CHECK: omptarget message: device mapping required by 'present' map type modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]] bytes)
-  // CHECK: omptarget error: Call to getTargetPointer returned null pointer ('present' map type modifier).
-  // CHECK: omptarget error: Call to targetDataBegin failed, abort target.
-  // CHECK: omptarget error: Failed to process data before launching the kernel.
-  // CHECK: omptarget fatal error 1: failure of target construct while offloading is mandatory
+  // CHECK: omptarget message: device mapping required by 'present' map type
+  // modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]]
+  // bytes) CHECK: omptarget error: Call to getTargetPointer returned null
+  // pointer ('present' map type modifier). CHECK: omptarget error: Call to
+  // targetDataBegin failed, abort target. CHECK: omptarget error: Failed to
+  // process data before launching the kernel. CHECK: omptarget fatal error 1:
+  // failure of target construct while offloading is mandatory
 #pragma omp target map(present, alloc : i)
   ;
 
diff --git a/offload/test/mapping/present/target_array_extension.c b/offload/test/mapping/present/target_array_extension.c
index 873b2b3617..132a2430ff 100644
--- a/offload/test/mapping/present/target_array_extension.c
+++ b/offload/test/mapping/present/target_array_extension.c
@@ -42,12 +42,8 @@
 #define SMALL_SIZE (SMALL_END - SMALL_BEG)
 #define LARGE_SIZE (LARGE_END - LARGE_BEG)
 
-#define SMALL                                                                  \
-  SMALL_BEG:                                                                   \
-  SMALL_SIZE
-#define LARGE                                                                  \
-  LARGE_BEG:                                                                   \
-  LARGE_SIZE
+#define SMALL SMALL_BEG : SMALL_SIZE
+#define LARGE LARGE_BEG : LARGE_SIZE
 
 int main() {
   int arr[SIZE];
@@ -70,12 +66,16 @@ int main() {
   // CHECK: arr is present
   fprintf(stderr, "arr is present\n");
 
-  // CHECK: omptarget message: explicit extension not allowed: host address specified is 0x{{0*}}[[#LARGE_ADDR]] ([[#LARGE_BYTES]] bytes), but device allocation maps to host at 0x{{0*}}[[#SMALL_ADDR]] ([[#SMALL_BYTES]] bytes)
-  // CHECK: omptarget message: device mapping required by 'present' map type modifier does not exist for host address 0x{{0*}}[[#LARGE_ADDR]] ([[#LARGE_BYTES]] bytes)
-  // CHECK: omptarget error: Call to getTargetPointer returned null pointer ('present' map type modifier).
-  // CHECK: omptarget error: Call to targetDataBegin failed, abort target.
-  // CHECK: omptarget error: Failed to process data before launching the kernel.
-  // CHECK: omptarget fatal error 1: failure of target construct while offloading is mandatory
+  // CHECK: omptarget message: explicit extension not allowed: host address
+  // specified is 0x{{0*}}[[#LARGE_ADDR]] ([[#LARGE_BYTES]] bytes), but device
+  // allocation maps to host at 0x{{0*}}[[#SMALL_ADDR]] ([[#SMALL_BYTES]] bytes)
+  // CHECK: omptarget message: device mapping required by 'present' map type
+  // modifier does not exist for host address 0x{{0*}}[[#LARGE_ADDR]]
+  // ([[#LARGE_BYTES]] bytes) CHECK: omptarget error: Call to getTargetPointer
+  // returned null pointer ('present' map type modifier). CHECK: omptarget
+  // error: Call to targetDataBegin failed, abort target. CHECK: omptarget
+  // error: Failed to process data before launching the kernel. CHECK: omptarget
+  // fatal error 1: failure of target construct while offloading is mandatory
 #pragma omp target data map(alloc : arr[SMALL])
   {
 #pragma omp target map(present, tofrom : arr[LARGE])
diff --git a/offload/test/mapping/present/target_data.c b/offload/test/mapping/present/target_data.c
index f894283a0c..ae3b762c34 100644
--- a/offload/test/mapping/present/target_data.c
+++ b/offload/test/mapping/present/target_data.c
@@ -18,8 +18,10 @@ int main() {
   // CHECK: i is present
   fprintf(stderr, "i is present\n");
 
-  // CHECK: omptarget message: device mapping required by 'present' map type modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]] bytes)
-  // CHECK: omptarget fatal error 1: failure of target construct while offloading is mandatory
+  // CHECK: omptarget message: device mapping required by 'present' map type
+  // modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]]
+  // bytes) CHECK: omptarget fatal error 1: failure of target construct while
+  // offloading is mandatory
 #pragma omp target data map(present, alloc : i)
   ;
 
diff --git a/offload/test/mapping/present/target_data_array_extension.c b/offload/test/mapping/present/target_data_array_extension.c
index 794543a246..7124664756 100644
--- a/offload/test/mapping/present/target_data_array_extension.c
+++ b/offload/test/mapping/present/target_data_array_extension.c
@@ -42,12 +42,8 @@
 #define SMALL_SIZE (SMALL_END - SMALL_BEG)
 #define LARGE_SIZE (LARGE_END - LARGE_BEG)
 
-#define SMALL                                                                  \
-  SMALL_BEG:                                                                   \
-  SMALL_SIZE
-#define LARGE                                                                  \
-  LARGE_BEG:                                                                   \
-  LARGE_SIZE
+#define SMALL SMALL_BEG : SMALL_SIZE
+#define LARGE LARGE_BEG : LARGE_SIZE
 
 int main() {
   int arr[SIZE];
@@ -70,10 +66,14 @@ int main() {
   // CHECK: arr is present
   fprintf(stderr, "arr is present\n");
 
-  // CHECK: omptarget message: explicit extension not allowed: host address specified is 0x{{0*}}[[#LARGE_ADDR]] ([[#LARGE_BYTES]] bytes), but device allocation maps to host at 0x{{0*}}[[#SMALL_ADDR]] ([[#SMALL_BYTES]] bytes)
-  // CHECK: omptarget message: device mapping required by 'present' map type modifier does not exist for host address 0x{{0*}}[[#LARGE_ADDR]] ([[#LARGE_BYTES]] bytes)
-  // CHECK: omptarget error: Call to getTargetPointer returned null pointer ('present' map type modifier).
-  // CHECK: omptarget fatal error 1: failure of target construct while offloading is mandatory
+  // CHECK: omptarget message: explicit extension not allowed: host address
+  // specified is 0x{{0*}}[[#LARGE_ADDR]] ([[#LARGE_BYTES]] bytes), but device
+  // allocation maps to host at 0x{{0*}}[[#SMALL_ADDR]] ([[#SMALL_BYTES]] bytes)
+  // CHECK: omptarget message: device mapping required by 'present' map type
+  // modifier does not exist for host address 0x{{0*}}[[#LARGE_ADDR]]
+  // ([[#LARGE_BYTES]] bytes) CHECK: omptarget error: Call to getTargetPointer
+  // returned null pointer ('present' map type modifier). CHECK: omptarget fatal
+  // error 1: failure of target construct while offloading is mandatory
 #pragma omp target data map(alloc : arr[SMALL])
   {
 #pragma omp target data map(present, tofrom : arr[LARGE])
diff --git a/offload/test/mapping/present/target_enter_data.c b/offload/test/mapping/present/target_enter_data.c
index 871a05290e..633277ddf0 100644
--- a/offload/test/mapping/present/target_enter_data.c
+++ b/offload/test/mapping/present/target_enter_data.c
@@ -18,9 +18,11 @@ int main() {
   // CHECK: i is present
   fprintf(stderr, "i is present\n");
 
-  // CHECK: omptarget message: device mapping required by 'present' map type modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]] bytes)
-  // CHECK: omptarget error: Call to getTargetPointer returned null pointer ('present' map type modifier).
-  // CHECK: omptarget fatal error 1: failure of target construct while offloading is mandatory
+  // CHECK: omptarget message: device mapping required by 'present' map type
+  // modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]]
+  // bytes) CHECK: omptarget error: Call to getTargetPointer returned null
+  // pointer ('present' map type modifier). CHECK: omptarget fatal error 1:
+  // failure of target construct while offloading is mandatory
 #pragma omp target enter data map(present, alloc : i)
 
   // CHECK-NOT: i is present
diff --git a/offload/test/mapping/present/target_exit_data_delete.c b/offload/test/mapping/present/target_exit_data_delete.c
index 0fb812b299..d791166147 100644
--- a/offload/test/mapping/present/target_exit_data_delete.c
+++ b/offload/test/mapping/present/target_exit_data_delete.c
@@ -17,8 +17,10 @@ int main() {
   // CHECK: i was present
   fprintf(stderr, "i was present\n");
 
-// CHECK: omptarget message: device mapping required by 'present' map type modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]] bytes)
-// CHECK: omptarget fatal error 1: failure of target construct while offloading is mandatory
+// CHECK: omptarget message: device mapping required by 'present' map type
+// modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]]
+// bytes) CHECK: omptarget fatal error 1: failure of target construct while
+// offloading is mandatory
 #pragma omp target exit data map(present, delete : i)
 
   // CHECK-NOT: i was present
diff --git a/offload/test/mapping/present/target_exit_data_release.c b/offload/test/mapping/present/target_exit_data_release.c
index 14be22faba..083f6c080d 100644
--- a/offload/test/mapping/present/target_exit_data_release.c
+++ b/offload/test/mapping/present/target_exit_data_release.c
@@ -17,8 +17,10 @@ int main() {
   // CHECK: i was present
   fprintf(stderr, "i was present\n");
 
-// CHECK: omptarget message: device mapping required by 'present' map type modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]] bytes)
-// CHECK: omptarget fatal error 1: failure of target construct while offloading is mandatory
+// CHECK: omptarget message: device mapping required by 'present' map type
+// modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]]
+// bytes) CHECK: omptarget fatal error 1: failure of target construct while
+// offloading is mandatory
 #pragma omp target exit data map(present, release : i)
 
   // CHECK-NOT: i was present
diff --git a/offload/test/mapping/present/target_update.c b/offload/test/mapping/present/target_update.c
index 9f6783b6ef..5df4eee6fc 100644
--- a/offload/test/mapping/present/target_update.c
+++ b/offload/test/mapping/present/target_update.c
@@ -32,8 +32,10 @@ int main() {
   // CHECK: i is present
   fprintf(stderr, "i is present\n");
 
-  // CHECK: omptarget message: device mapping required by 'present' motion modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]] bytes)
-  // CHECK: omptarget fatal error 1: failure of target construct while offloading is mandatory
+  // CHECK: omptarget message: device mapping required by 'present' motion
+  // modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]]
+  // bytes) CHECK: omptarget fatal error 1: failure of target construct while
+  // offloading is mandatory
 #pragma omp target update CLAUSE(present : i)
 
   // CHECK-NOT: i is present
diff --git a/offload/test/mapping/present/target_update_array_extension.c b/offload/test/mapping/present/target_update_array_extension.c
index 11ad4a8d49..f8f3c94d21 100644
--- a/offload/test/mapping/present/target_update_array_extension.c
+++ b/offload/test/mapping/present/target_update_array_extension.c
@@ -66,8 +66,10 @@ int main() {
   // CHECK: arr is present
   fprintf(stderr, "arr is present\n");
 
-  // CHECK: omptarget message: device mapping required by 'present' motion modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]] bytes)
-  // CHECK: omptarget fatal error 1: failure of target construct while offloading is mandatory
+  // CHECK: omptarget message: device mapping required by 'present' motion
+  // modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]]
+  // bytes) CHECK: omptarget fatal error 1: failure of target construct while
+  // offloading is mandatory
 #pragma omp target data map(alloc : arr[SMALL])
   {
 #pragma omp target update CLAUSE(present : arr[LARGE])
diff --git a/offload/test/mapping/present/zero_length_array_section.c b/offload/test/mapping/present/zero_length_array_section.c
index e903a268ef..4759585845 100644
--- a/offload/test/mapping/present/zero_length_array_section.c
+++ b/offload/test/mapping/present/zero_length_array_section.c
@@ -20,11 +20,13 @@ int main() {
 
   // arr[0:0] doesn't create an actual mapping in the first directive.
   //
-  // CHECK: omptarget message: device mapping required by 'present' map type modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] (0 bytes)
-  // CHECK: omptarget error: Call to getTargetPointer returned null pointer ('present' map type modifier).
-  // CHECK: omptarget error: Call to targetDataBegin failed, abort target.
-  // CHECK: omptarget error: Failed to process data before launching the kernel.
-  // CHECK: omptarget fatal error 1: failure of target construct while offloading is mandatory
+  // CHECK: omptarget message: device mapping required by 'present' map type
+  // modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] (0 bytes)
+  // CHECK: omptarget error: Call to getTargetPointer returned null pointer
+  // ('present' map type modifier). CHECK: omptarget error: Call to
+  // targetDataBegin failed, abort target. CHECK: omptarget error: Failed to
+  // process data before launching the kernel. CHECK: omptarget fatal error 1:
+  // failure of target construct while offloading is mandatory
 #pragma omp target data map(alloc : arr[0 : 0])
 #pragma omp target map(present, alloc : arr[0 : 0])
   ;
diff --git a/offload/test/mapping/present/zero_length_array_section_exit.c b/offload/test/mapping/present/zero_length_array_section_exit.c
index 5a7360542e..c88c47f115 100644
--- a/offload/test/mapping/present/zero_length_array_section_exit.c
+++ b/offload/test/mapping/present/zero_length_array_section_exit.c
@@ -19,8 +19,10 @@ int main() {
 
   // arr[0:0] doesn't create an actual mapping in the first directive.
   //
-  // CHECK: omptarget message: device mapping required by 'present' map type modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] (0 bytes)
-  // CHECK: omptarget fatal error 1: failure of target construct while offloading is mandatory
+  // CHECK: omptarget message: device mapping required by 'present' map type
+  // modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] (0 bytes)
+  // CHECK: omptarget fatal error 1: failure of target construct while
+  // offloading is mandatory
 #pragma omp target enter data map(alloc : arr[0 : 0])
 #pragma omp target exit data map(present, release : arr[0 : 0])
 
diff --git a/offload/test/mapping/private_mapping.c b/offload/test/mapping/private_mapping.c
index 1329a66a5d..157cc4c511 100644
--- a/offload/test/mapping/private_mapping.c
+++ b/offload/test/mapping/private_mapping.c
@@ -10,9 +10,10 @@ int main() {
   int data3[3] = {100, 200, 500};
   int sum[16] = {0};
 
-  for (int i=0; i<16; i++) sum[i] = 10000;
+  for (int i = 0; i < 16; i++)
+    sum[i] = 10000;
 
-#pragma omp target teams distribute parallel for map(tofrom : sum[:16])       \
+#pragma omp target teams distribute parallel for map(tofrom : sum[ : 16])      \
     firstprivate(data1, data2, data3)
   for (int i = 0; i < 16; ++i) {
     for (int j = 0; j < 3; ++j) {
diff --git a/offload/test/mapping/target_data_array_extension_at_exit.c b/offload/test/mapping/target_data_array_extension_at_exit.c
index e300c800b2..154dc66080 100644
--- a/offload/test/mapping/target_data_array_extension_at_exit.c
+++ b/offload/test/mapping/target_data_array_extension_at_exit.c
@@ -39,12 +39,8 @@
 #error EXTENDS undefined
 #endif
 
-#define SMALL                                                                  \
-  SMALL_BEG:                                                                   \
-  (SMALL_END - SMALL_BEG)
-#define LARGE                                                                  \
-  LARGE_BEG:                                                                   \
-  (LARGE_END - LARGE_BEG)
+#define SMALL SMALL_BEG : (SMALL_END - SMALL_BEG)
+#define LARGE LARGE_BEG : (LARGE_END - LARGE_BEG)
 
 void check_not_present() {
   int arr[SIZE];
diff --git a/offload/test/mapping/target_implicit_partial_map.c b/offload/test/mapping/target_implicit_partial_map.c
index a8b2cc893d..fe4b9df72c 100644
--- a/offload/test/mapping/target_implicit_partial_map.c
+++ b/offload/test/mapping/target_implicit_partial_map.c
@@ -29,7 +29,7 @@ int main() {
       arr[50] = 5;
       arr[51] = 6;
     } // must treat as present (dec ref count) even though full size not present
-  }   // wouldn't delete if previous ref count dec didn't happen
+  } // wouldn't delete if previous ref count dec didn't happen
 
   // CHECK: arr[50] still present: 0
   fprintf(stderr, "arr[50] still present: %d\n",
diff --git a/offload/test/mapping/target_wrong_use_device_addr.c b/offload/test/mapping/target_wrong_use_device_addr.c
index 7a5babd692..28ec6857fa 100644
--- a/offload/test/mapping/target_wrong_use_device_addr.c
+++ b/offload/test/mapping/target_wrong_use_device_addr.c
@@ -14,7 +14,7 @@ int main() {
   // CHECK: host addr=0x[[#%x,HOST_ADDR:]]
   fprintf(stderr, "host addr=%p\n", x);
 
-#pragma omp target data map(to : x [0:10])
+#pragma omp target data map(to : x[0 : 10])
   {
 // CHECK: omptarget device 0 info: variable x does not have a valid device
 // counterpart
@@ -27,4 +27,3 @@ int main() {
 
   return 0;
 }
-
diff --git a/offload/test/offloading/back2back_distribute.c b/offload/test/offloading/back2back_distribute.c
index 63cabd0c66..36d2512deb 100644
--- a/offload/test/offloading/back2back_distribute.c
+++ b/offload/test/offloading/back2back_distribute.c
@@ -1,4 +1,5 @@
-// RUN: %libomptarget-compile-generic -O3 && %libomptarget-run-generic | %fcheck-generic
+// RUN: %libomptarget-compile-generic -O3 && %libomptarget-run-generic |
+// %fcheck-generic
 
 #include <omp.h>
 #include <stdio.h>
@@ -7,10 +8,10 @@
 #define MAX_N 25000
 
 void reset_input(double *a, double *a_h, double *b, double *c) {
-  for(int i = 0 ; i < MAX_N ; i++) {
+  for (int i = 0; i < MAX_N; i++) {
     a[i] = a_h[i] = i;
-    b[i] = i*2;
-    c[i] = i-3;
+    b[i] = i * 2;
+    c[i] = i - 3;
   }
 }
 
@@ -22,15 +23,16 @@ int main(int argc, char *argv[]) {
   double *b = (double *)calloc(MAX_N, sizeof(double));
   double *c = (double *)calloc(MAX_N, sizeof(double));
 
-#pragma omp target enter data map(to:a[:MAX_N],b[:MAX_N],c[:MAX_N],d[:MAX_N])
+#pragma omp target enter data map(to : a[ : MAX_N], b[ : MAX_N], c[ : MAX_N],  \
+                                      d[ : MAX_N])
 
-  for (int n = 32 ; n < MAX_N ; n+=5000) {
+  for (int n = 32; n < MAX_N; n += 5000) {
     reset_input(a, a_h, b, c);
 
-#pragma omp target update to(a[:n],b[:n],c[:n],d[:n])
+#pragma omp target update to(a[ : n], b[ : n], c[ : n], d[ : n])
     int t = 0;
-    for (int tms = 1 ; tms <= 256 ; tms *= 2) { // 8 times
-      for (int ths = 32 ; ths <= 1024 ; ths *= 2) { // 6 times
+    for (int tms = 1; tms <= 256; tms *= 2) {     // 8 times
+      for (int ths = 32; ths <= 1024; ths *= 2) { // 6 times
         t++;
 #pragma omp target
 #pragma omp teams num_teams(tms) thread_limit(ths)
@@ -41,29 +43,31 @@ int main(int argc, char *argv[]) {
           }
 #pragma omp distribute parallel for
           for (int i = 0; i < n; ++i) {
-           d[i] -= b[i] + c[i];
+            d[i] -= b[i] + c[i];
           }
         }
       } // loop over 'ths'
     } // loop over 'tms'
 
     // check results for each 'n'
-    for (int times = 0 ; times < t ; times++) {
+    for (int times = 0; times < t; times++) {
       for (int i = 0; i < n; ++i) {
         a_h[i] += b[i] + c[i];
       }
       for (int i = 0; i < n; ++i)
         d_h[i] -= b[i] + c[i];
     }
-#pragma omp target update from(a[:n],d[:n])
+#pragma omp target update from(a[ : n], d[ : n])
 
     for (int i = 0; i < n; ++i) {
       if (a_h[i] != a[i]) {
-        printf("A Error at n = %d, i = %d: host = %f, device = %f\n", n, i, a_h[i], a[i]);
+        printf("A Error at n = %d, i = %d: host = %f, device = %f\n", n, i,
+               a_h[i], a[i]);
         return 1;
       }
       if (d_h[i] != d[i]) {
-        printf("D Error at n = %d, i = %d: host = %lf, device = %lf\n", n, i, d_h[i], d[i]);
+        printf("D Error at n = %d, i = %d: host = %lf, device = %lf\n", n, i,
+               d_h[i], d[i]);
         return 1;
       }
     }
diff --git a/offload/test/offloading/bug49021.cpp b/offload/test/offloading/bug49021.cpp
index 37da8ce5d8..8b03c4580e 100644
--- a/offload/test/offloading/bug49021.cpp
+++ b/offload/test/offloading/bug49021.cpp
@@ -1,7 +1,8 @@
 // RUN: %libomptarget-compilexx-generic -O3 && %libomptarget-run-generic
-// RUN: %libomptarget-compilexx-generic -O3 -ffast-math && %libomptarget-run-generic
-// RUN: %libomptarget-compileoptxx-generic -O3 && %libomptarget-run-generic
-// RUN: %libomptarget-compileoptxx-generic -O3 -ffast-math && %libomptarget-run-generic
+// RUN: %libomptarget-compilexx-generic -O3 -ffast-math &&
+// %libomptarget-run-generic RUN: %libomptarget-compileoptxx-generic -O3 &&
+// %libomptarget-run-generic RUN: %libomptarget-compileoptxx-generic -O3
+// -ffast-math && %libomptarget-run-generic
 
 #include <iostream>
 
diff --git a/offload/test/offloading/bug49334.cpp b/offload/test/offloading/bug49334.cpp
index 1f19dab378..a8f6e814c5 100644
--- a/offload/test/offloading/bug49334.cpp
+++ b/offload/test/offloading/bug49334.cpp
@@ -93,7 +93,7 @@ int BlockMatMul_TargetNowait(BlockMatrix &A, BlockMatrix &B, BlockMatrix &C) {
       for (int k = 0; k < N / BS; ++k) {
         float *BlockA = A.GetBlock(i, k);
         float *BlockB = B.GetBlock(k, j);
-// clang-format off
+        // clang-format off
 #pragma omp target depend(in: BlockA[0], BlockB[0]) depend(inout: BlockC[0])   \
             map(to: BlockA[:BS * BS], BlockB[:BS * BS])                        \
             map(tofrom: BlockC[:BS * BS]) nowait
diff --git a/offload/test/offloading/complex_reduction.cpp b/offload/test/offloading/complex_reduction.cpp
index 3239f48743..32f06ae4cb 100644
--- a/offload/test/offloading/complex_reduction.cpp
+++ b/offload/test/offloading/complex_reduction.cpp
@@ -1,5 +1,6 @@
 // RUN: %libomptarget-compilexx-generic -O3 && %libomptarget-run-generic
-// RUN: %libomptarget-compilexx-generic -O3 -ffast-math && %libomptarget-run-generic
+// RUN: %libomptarget-compilexx-generic -O3 -ffast-math &&
+// %libomptarget-run-generic
 
 #include <complex>
 #include <iostream>
@@ -20,8 +21,8 @@ template <typename T> void test_map() {
 }
 
 #if !defined(__NO_UDR)
-#pragma omp declare reduction(+ : std::complex <float> : omp_out += omp_in)
-#pragma omp declare reduction(+ : std::complex <double> : omp_out += omp_in)
+#pragma omp declare reduction(+ : std::complex<float> : omp_out += omp_in)
+#pragma omp declare reduction(+ : std::complex<double> : omp_out += omp_in)
 #endif
 
 template <typename T> class initiator {
@@ -43,7 +44,8 @@ template <typename T> void test_reduction() {
     sum_host += array[i];
   }
 
-#pragma omp target teams distribute parallel for map(to : array[:size]) reduction(+ : sum)
+#pragma omp target teams distribute parallel for map(to : array[ : size])      \
+    reduction(+ : sum)
   for (int i = 0; i < size; i++)
     sum += array[i];
 
@@ -55,10 +57,8 @@ template <typename T> void test_reduction() {
 
   const int nblock(10), block_size(10);
   T block_sum[nblock];
-#pragma omp target teams distribute map(to                                     \
-                                        : array[:size])                        \
-    map(from                                                                   \
-        : block_sum[:nblock])
+#pragma omp target teams distribute map(to : array[ : size])                   \
+    map(from : block_sum[ : nblock])
   for (int ib = 0; ib < nblock; ib++) {
     T partial_sum(0);
     const int istart = ib * block_size;
diff --git a/offload/test/offloading/dynamic_module_load.c b/offload/test/offloading/dynamic_module_load.c
index 5393f33e84..6e39669021 100644
--- a/offload/test/offloading/dynamic_module_load.c
+++ b/offload/test/offloading/dynamic_module_load.c
@@ -1,4 +1,6 @@
-// RUN: %libomptarget-compile-generic -DSHARED -fPIC -shared -o %t.so && %clang %flags %s -o %t -ldl && %libomptarget-run-generic %t.so 2>&1 | %fcheck-generic
+// RUN: %libomptarget-compile-generic -DSHARED -fPIC -shared -o %t.so && %clang
+// %flags %s -o %t -ldl && %libomptarget-run-generic %t.so 2>&1 |
+// %fcheck-generic
 
 #ifdef SHARED
 #include <stdio.h>
diff --git a/offload/test/offloading/extern.c b/offload/test/offloading/extern.c
index 35cf905ee0..df9285ce61 100644
--- a/offload/test/offloading/extern.c
+++ b/offload/test/offloading/extern.c
@@ -1,11 +1,12 @@
 // RUN: %libomptarget-compile-generic -DVAR -c -o %t.o
-// RUN: %libomptarget-compile-generic %t.o && %libomptarget-run-generic | %fcheck-generic
+// RUN: %libomptarget-compile-generic %t.o && %libomptarget-run-generic |
+// %fcheck-generic
 
 #ifdef VAR
 int x = 1;
 #else
-#include <stdio.h>
 #include <assert.h>
+#include <stdio.h>
 extern int x;
 
 int main() {
@@ -22,6 +23,6 @@ int main() {
   assert(value == 999);
 
   // CHECK: PASS
-  printf ("PASS\n");
+  printf("PASS\n");
 }
 #endif
diff --git a/offload/test/offloading/fortran/basic_array.c b/offload/test/offloading/fortran/basic_array.c
index 4daf1d4b0a..0928814bdd 100644
--- a/offload/test/offloading/fortran/basic_array.c
+++ b/offload/test/offloading/fortran/basic_array.c
@@ -14,7 +14,7 @@ void increment_at(int i, int *array);
 #pragma omp end declare target
 
 void increment_array(int *b, int n) {
-#pragma omp target map(tofrom : b [0:n])
+#pragma omp target map(tofrom : b[0 : n])
   for (int i = 0; i < n; i++) {
     increment_at(i, b);
   }
diff --git a/offload/test/offloading/global_constructor.cpp b/offload/test/offloading/global_constructor.cpp
index eb68c5f783..6260d8994c 100644
--- a/offload/test/offloading/global_constructor.cpp
+++ b/offload/test/offloading/global_constructor.cpp
@@ -1,4 +1,5 @@
-// RUN: %libomptarget-compilexx-generic && %libomptarget-run-generic | %fcheck-generic
+// RUN: %libomptarget-compilexx-generic && %libomptarget-run-generic |
+// %fcheck-generic
 
 #include <cstdio>
 
diff --git a/offload/test/offloading/info.c b/offload/test/offloading/info.c
index da8e4c44c5..5e63a252e5 100644
--- a/offload/test/offloading/info.c
+++ b/offload/test/offloading/info.c
@@ -25,7 +25,7 @@ int main() {
   int C[N];
   int val = 1;
 
-// clang-format off
+  // clang-format off
 // INFO: info: Entering OpenMP data region with being_mapper at info.c:{{[0-9]+}}:{{[0-9]+}} with 3 arguments:
 // INFO: info: alloc(A[0:64])[256]
 // INFO: info: tofrom(B[0:64])[256]
@@ -60,7 +60,7 @@ int main() {
 // INFO: info: OpenMP Host-Device pointer mappings after block at info.c:[[#%u,]]:[[#%u,]]:
 // INFO: info: Host Ptr  Target Ptr Size (B) DynRefCount HoldRefCount Declaration
 // INFO: info: [[#%#x,]] [[#%#x,]]  4        INF         0            global at unknown:0:0
-// clang-format on
+  // clang-format on
 #pragma omp target data map(alloc : A[0 : N])                                  \
     map(ompx_hold, tofrom : B[0 : N]) map(to : C[0 : N])
 #pragma omp target firstprivate(val)
diff --git a/offload/test/offloading/looptripcnt.c b/offload/test/offloading/looptripcnt.c
index d1a095038b..5547bbac16 100644
--- a/offload/test/offloading/looptripcnt.c
+++ b/offload/test/offloading/looptripcnt.c
@@ -1,5 +1,6 @@
-// RUN: %libomptarget-compile-generic && env LIBOMPTARGET_DEBUG=1 %libomptarget-run-generic 2>&1 | %fcheck-generic -allow-empty -check-prefix=DEBUG
-// REQUIRES: libomptarget-debug
+// RUN: %libomptarget-compile-generic && env LIBOMPTARGET_DEBUG=1
+// %libomptarget-run-generic 2>&1 | %fcheck-generic -allow-empty
+// -check-prefix=DEBUG REQUIRES: libomptarget-debug
 
 /*
   Test for looptripcount being popped from runtime stack.
diff --git a/offload/test/offloading/mandatory_but_no_devices.c b/offload/test/offloading/mandatory_but_no_devices.c
index ecdee72aca..c44b3f6ba8 100644
--- a/offload/test/offloading/mandatory_but_no_devices.c
+++ b/offload/test/offloading/mandatory_but_no_devices.c
@@ -47,7 +47,8 @@
 #include <omp.h>
 #include <stdio.h>
 
-// CHECK: omptarget fatal error 1: failure of target construct while offloading is mandatory
+// CHECK: omptarget fatal error 1: failure of target construct while offloading
+// is mandatory
 int main(void) {
   int X;
 #pragma omp DIR device(omp_get_initial_device())
diff --git a/offload/test/offloading/non_contiguous_update.cpp b/offload/test/offloading/non_contiguous_update.cpp
index 609f0f967f..3d9ce82361 100644
--- a/offload/test/offloading/non_contiguous_update.cpp
+++ b/offload/test/offloading/non_contiguous_update.cpp
@@ -1,5 +1,6 @@
-// RUN: %libomptarget-compile-generic && env LIBOMPTARGET_DEBUG=1 %libomptarget-run-generic 2>&1 | %fcheck-generic -allow-empty -check-prefix=DEBUG
-// REQUIRES: libomptarget-debug
+// RUN: %libomptarget-compile-generic && env LIBOMPTARGET_DEBUG=1
+// %libomptarget-run-generic 2>&1 | %fcheck-generic -allow-empty
+// -check-prefix=DEBUG REQUIRES: libomptarget-debug
 
 #include <cassert>
 #include <cstdio>
diff --git a/offload/test/offloading/ompx_bare.c b/offload/test/offloading/ompx_bare.c
index 3dabdcd15e..991672a21a 100644
--- a/offload/test/offloading/ompx_bare.c
+++ b/offload/test/offloading/ompx_bare.c
@@ -20,9 +20,11 @@ int main(int argc, char *argv[]) {
   const int N = num_blocks * block_size;
   int *data = (int *)malloc(N * sizeof(int));
 
-  // CHECK: "PluginInterface" device 0 info: Launching kernel __omp_offloading_{{.*}} with 64 blocks and 64 threads in SPMD mode
+  // CHECK: "PluginInterface" device 0 info: Launching kernel
+  // __omp_offloading_{{.*}} with 64 blocks and 64 threads in SPMD mode
 
-#pragma omp target teams ompx_bare num_teams(num_blocks) thread_limit(block_size) map(from: data[0:N])
+#pragma omp target teams ompx_bare num_teams(num_blocks)                       \
+    thread_limit(block_size) map(from : data[0 : N])
   {
     int bid = ompx_block_id_x();
     int bdim = ompx_block_dim_x();
diff --git a/offload/test/offloading/ompx_coords.c b/offload/test/offloading/ompx_coords.c
index 5e4e14b4c6..7a6eb27351 100644
--- a/offload/test/offloading/ompx_coords.c
+++ b/offload/test/offloading/ompx_coords.c
@@ -28,7 +28,7 @@ int main(int argc, char **argv) {
   int TL = 256;
   int NT = (N + TL - 1) / TL;
 
-#pragma omp target data map(tofrom : X [0:N])
+#pragma omp target data map(tofrom : X[0 : N])
 #pragma omp target teams num_teams(NT) thread_limit(TL)
   {
 #pragma omp parallel
diff --git a/offload/test/offloading/ompx_saxpy_mixed.c b/offload/test/offloading/ompx_saxpy_mixed.c
index f479be8a48..b55c7e7341 100644
--- a/offload/test/offloading/ompx_saxpy_mixed.c
+++ b/offload/test/offloading/ompx_saxpy_mixed.c
@@ -30,7 +30,7 @@ int main(int argc, char **argv) {
   int TL = 256;
   int NT = (N + TL - 1) / TL;
 
-#pragma omp target data map(to : X [0:N]) map(Y [0:N])
+#pragma omp target data map(to : X[0 : N]) map(Y[0 : N])
 #pragma omp target teams num_teams(NT) thread_limit(TL)
   {
 #pragma omp parallel
diff --git a/offload/test/offloading/spmdization.c b/offload/test/offloading/spmdization.c
index 77913bec83..26bd7263c0 100644
--- a/offload/test/offloading/spmdization.c
+++ b/offload/test/offloading/spmdization.c
@@ -31,14 +31,16 @@ int main(void) {
     nested = omp_get_nested();
     tid = omp_get_thread_num();
     maxt = omp_get_max_threads();
-    #pragma omp parallel
-    noop();  
+#pragma omp parallel
+    noop();
   }
-  printf("NumThreads: %i, InParallel: %i, Level: %i, ActiveLevel: %i, Nested: %i, "
-         "ThreadNum: %i, MaxThreads: %i\n",
-         nthreads, ip, lvl, alvl, nested, tid, maxt);
+  printf(
+      "NumThreads: %i, InParallel: %i, Level: %i, ActiveLevel: %i, Nested: %i, "
+      "ThreadNum: %i, MaxThreads: %i\n",
+      nthreads, ip, lvl, alvl, nested, tid, maxt);
   // GENERIC: Generic mode
   // SPMD: Generic-SPMD mode
-  // CHECK: NumThreads: 1, InParallel: 0, Level: 0, ActiveLevel: 0, Nested: 0, ThreadNum: 0, MaxThreads: 
+  // CHECK: NumThreads: 1, InParallel: 0, Level: 0, ActiveLevel: 0, Nested: 0,
+  // ThreadNum: 0, MaxThreads:
   return 0;
 }
diff --git a/offload/test/offloading/static_linking.c b/offload/test/offloading/static_linking.c
index 7be95a10ff..864ec4f294 100644
--- a/offload/test/offloading/static_linking.c
+++ b/offload/test/offloading/static_linking.c
@@ -1,6 +1,7 @@
 // RUN: %libomptarget-compile-generic -DLIBRARY -c -o %t.o
 // RUN: ar rcs %t.a %t.o
-// RUN: %libomptarget-compile-generic %t.a && %libomptarget-run-generic 2>&1 | %fcheck-generic
+// RUN: %libomptarget-compile-generic %t.a && %libomptarget-run-generic 2>&1 |
+// %fcheck-generic
 
 #ifdef LIBRARY
 int x = 42;
diff --git a/offload/test/offloading/target-tile.c b/offload/test/offloading/target-tile.c
index 8460b43b6f..2e67ed8ede 100644
--- a/offload/test/offloading/target-tile.c
+++ b/offload/test/offloading/target-tile.c
@@ -14,10 +14,10 @@
 int main() {
   int order[I_NTILES][J_NTILES][I_NELEMS][J_NELEMS];
   int i, j;
-  #pragma omp target map(tofrom: i, j)
+#pragma omp target map(tofrom : i, j)
   {
     int next = 0;
-    #pragma omp tile sizes(I_NELEMS, J_NELEMS)
+#pragma omp tile sizes(I_NELEMS, J_NELEMS)
     for (i = 0; i < I_NTILES * I_NELEMS; ++i) {
       for (j = 0; j < J_NTILES * J_NELEMS; ++j) {
         int iTile = i / I_NELEMS;
@@ -35,8 +35,8 @@ int main() {
         for (int jElem = 0; jElem < J_NELEMS; ++jElem) {
           int actual = order[iTile][jTile][iElem][jElem];
           if (expected != actual) {
-            printf("error: order[%d][%d][%d][%d] = %d, expected %d\n",
-                   iTile, jTile, iElem, jElem, actual, expected);
+            printf("error: order[%d][%d][%d][%d] = %d, expected %d\n", iTile,
+                   jTile, iElem, jElem, actual, expected);
             return 1;
           }
           ++expected;
diff --git a/offload/test/offloading/target_constexpr_mapping.cpp b/offload/test/offloading/target_constexpr_mapping.cpp
index 14cf92a7cc..3148b64fb0 100644
--- a/offload/test/offloading/target_constexpr_mapping.cpp
+++ b/offload/test/offloading/target_constexpr_mapping.cpp
@@ -18,7 +18,7 @@ constexpr static double anotherPi = 3.14;
 
 int main() {
   double a[2];
-#pragma omp target map(tofrom : a[:2])
+#pragma omp target map(tofrom : a[ : 2])
   {
     a[0] = A::pi;
     a[1] = anotherPi;
diff --git a/offload/test/offloading/target_critical_region.cpp b/offload/test/offloading/target_critical_region.cpp
index 495632bf76..4f513f7686 100644
--- a/offload/test/offloading/target_critical_region.cpp
+++ b/offload/test/offloading/target_critical_region.cpp
@@ -24,9 +24,7 @@ int main() {
   sum[0] = 0;
 
 #pragma omp target teams distribute parallel for num_teams(256)                \
-    schedule(static, 1) map(to                                                 \
-                            : A[:N]) map(tofrom                                \
-                                         : sum[:1])
+    schedule(static, 1) map(to : A[ : N]) map(tofrom : sum[ : 1])
   {
     for (int i = 0; i < N; i++) {
 #pragma omp critical
diff --git a/offload/test/offloading/target_depend_nowait.cpp b/offload/test/offloading/target_depend_nowait.cpp
index b0d5408770..f728ec5c79 100644
--- a/offload/test/offloading/target_depend_nowait.cpp
+++ b/offload/test/offloading/target_depend_nowait.cpp
@@ -43,7 +43,7 @@ int main() {
 #pragma omp target exit data map(release : A) map(from : B) depend(out : A[0]) \
     nowait
     } // if
-  }   // parallel
+  } // parallel
 
   int Sum = 0;
   for (int i = 0; i < N; i++)
diff --git a/offload/test/ompt/veccopy_disallow_both.c b/offload/test/ompt/veccopy_disallow_both.c
index a011c21955..0cf46c123a 100644
--- a/offload/test/ompt/veccopy_disallow_both.c
+++ b/offload/test/ompt/veccopy_disallow_both.c
@@ -72,35 +72,27 @@ int main() {
 /// CHECK-NOT: dest=(nil)
 /// CHECK: Callback DataOp EMI: endpoint=1 optype=2
 /// CHECK: Callback DataOp EMI: endpoint=2 optype=2
-/// CHECK: Callback Submit: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=1
-/// CHECK: Callback DataOp EMI: endpoint=1 optype=3
-/// CHECK: Callback DataOp EMI: endpoint=2 optype=3
-/// CHECK: Callback DataOp EMI: endpoint=1 optype=3
-/// CHECK: Callback DataOp EMI: endpoint=2 optype=3
-/// CHECK: Callback DataOp EMI: endpoint=1 optype=4
-/// CHECK: Callback DataOp EMI: endpoint=2 optype=4
-/// CHECK: Callback DataOp EMI: endpoint=1 optype=4
-/// CHECK: Callback DataOp EMI: endpoint=2 optype=4
-/// CHECK: Callback Target EMI: kind=1 endpoint=2
-/// CHECK: Callback Target EMI: kind=1 endpoint=1
-/// CHECK: Callback DataOp EMI: endpoint=1 optype=1
-/// CHECK: Callback DataOp EMI: endpoint=2 optype=1
-/// CHECK-NOT: dest=(nil)
-/// CHECK: Callback DataOp EMI: endpoint=1 optype=2
-/// CHECK: Callback DataOp EMI: endpoint=2 optype=2
+/// CHECK: Callback Submit: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=1 CHECK: Callback DataOp EMI:
+/// endpoint=1 optype=3 CHECK: Callback DataOp EMI: endpoint=2 optype=3 CHECK:
+/// Callback DataOp EMI: endpoint=1 optype=3 CHECK: Callback DataOp EMI:
+/// endpoint=2 optype=3 CHECK: Callback DataOp EMI: endpoint=1 optype=4 CHECK:
+/// Callback DataOp EMI: endpoint=2 optype=4 CHECK: Callback DataOp EMI:
+/// endpoint=1 optype=4 CHECK: Callback DataOp EMI: endpoint=2 optype=4 CHECK:
+/// Callback Target EMI: kind=1 endpoint=2 CHECK: Callback Target EMI: kind=1
+/// endpoint=1 CHECK: Callback DataOp EMI: endpoint=1 optype=1 CHECK: Callback
+/// DataOp EMI: endpoint=2 optype=1 CHECK-NOT: dest=(nil) CHECK: Callback DataOp
+/// EMI: endpoint=1 optype=2 CHECK: Callback DataOp EMI: endpoint=2 optype=2
 /// CHECK: Callback DataOp EMI: endpoint=1 optype=1
 /// CHECK: Callback DataOp EMI: endpoint=2 optype=1
 /// CHECK-NOT: dest=(nil)
 /// CHECK: Callback DataOp EMI: endpoint=1 optype=2
 /// CHECK: Callback DataOp EMI: endpoint=2 optype=2
-/// CHECK: Callback Submit: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=0
-/// CHECK: Callback DataOp EMI: endpoint=1 optype=3
-/// CHECK: Callback DataOp EMI: endpoint=2 optype=3
-/// CHECK: Callback DataOp EMI: endpoint=1 optype=3
-/// CHECK: Callback DataOp EMI: endpoint=2 optype=3
-/// CHECK: Callback DataOp EMI: endpoint=1 optype=4
-/// CHECK: Callback DataOp EMI: endpoint=2 optype=4
-/// CHECK: Callback DataOp EMI: endpoint=1 optype=4
-/// CHECK: Callback DataOp EMI: endpoint=2 optype=4
-/// CHECK: Callback Target EMI: kind=1 endpoint=2
-/// CHECK: Callback Fini:
+/// CHECK: Callback Submit: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=0 CHECK: Callback DataOp EMI:
+/// endpoint=1 optype=3 CHECK: Callback DataOp EMI: endpoint=2 optype=3 CHECK:
+/// Callback DataOp EMI: endpoint=1 optype=3 CHECK: Callback DataOp EMI:
+/// endpoint=2 optype=3 CHECK: Callback DataOp EMI: endpoint=1 optype=4 CHECK:
+/// Callback DataOp EMI: endpoint=2 optype=4 CHECK: Callback DataOp EMI:
+/// endpoint=1 optype=4 CHECK: Callback DataOp EMI: endpoint=2 optype=4 CHECK:
+/// Callback Target EMI: kind=1 endpoint=2 CHECK: Callback Fini:
diff --git a/offload/test/ompt/veccopy_map.c b/offload/test/ompt/veccopy_map.c
index 56a0dd48f5..7eb76651a9 100644
--- a/offload/test/ompt/veccopy_map.c
+++ b/offload/test/ompt/veccopy_map.c
@@ -56,31 +56,39 @@ int main() {
   return rc;
 }
 
-
 /// CHECK: 0: Could not register callback 'ompt_callback_target_map'
 /// CHECK: Callback Init:
 /// CHECK: Callback Load:
 /// CHECK: Callback Target: target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=1
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
-/// CHECK: Callback Submit: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=1
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
-/// CHECK: Callback Target: target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=2
+/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1 CHECK: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
+/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1 CHECK: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
+/// CHECK: Callback Submit: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=1 CHECK: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
+/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3 CHECK: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
+/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4 CHECK: Callback Target:
+/// target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=2
 
 /// CHECK: Callback Target: target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=1
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
-/// CHECK: Callback Submit: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=0
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
-/// CHECK: Callback Target: target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=2
-/// CHECK: Callback Fini:
+/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1 CHECK: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
+/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1 CHECK: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
+/// CHECK: Callback Submit: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=0 CHECK: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
+/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3 CHECK: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
+/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4 CHECK: Callback Target:
+/// target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=2 CHECK: Callback Fini:
diff --git a/offload/test/ompt/veccopy_no_device_init.c b/offload/test/ompt/veccopy_no_device_init.c
index c1a03191c3..f8872e29b3 100644
--- a/offload/test/ompt/veccopy_no_device_init.c
+++ b/offload/test/ompt/veccopy_no_device_init.c
@@ -54,26 +54,35 @@ int main() {
 /// CHECK-NOT: Callback Init:
 /// CHECK-NOT: Callback Load:
 /// CHECK-NOT: Callback Target: target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=1
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
-/// CHECK-NOT: Callback Submit: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=1
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
-/// CHECK-NOT: Callback Target: target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=2
+/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1 CHECK-NOT: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
+/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1 CHECK-NOT: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
+/// CHECK-NOT: Callback Submit: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=1 CHECK-NOT: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
+/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3 CHECK-NOT: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
+/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4 CHECK-NOT: Callback Target:
+/// target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=2
 
 /// CHECK-NOT: Callback Target: target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=1
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
-/// CHECK-NOT: Callback Submit: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=0
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
-/// CHECK-NOT: Callback Target: target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=2
-/// CHECK-NOT: Callback Fini:
+/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1 CHECK-NOT: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
+/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1 CHECK-NOT: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
+/// CHECK-NOT: Callback Submit: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=0 CHECK-NOT: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
+/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3 CHECK-NOT: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
+/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4 CHECK-NOT: Callback Target:
+/// target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=2 CHECK-NOT: Callback Fini:
diff --git a/offload/test/ompt/veccopy_wrong_return.c b/offload/test/ompt/veccopy_wrong_return.c
index 2d07b4e1bf..d88ba65e10 100644
--- a/offload/test/ompt/veccopy_wrong_return.c
+++ b/offload/test/ompt/veccopy_wrong_return.c
@@ -54,26 +54,35 @@ int main() {
 /// CHECK-NOT: Callback Init:
 /// CHECK-NOT: Callback Load:
 /// CHECK-NOT: Callback Target: target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=1
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
-/// CHECK-NOT: Callback Submit: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=1
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
-/// CHECK-NOT: Callback Target: target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=2
+/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1 CHECK-NOT: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
+/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1 CHECK-NOT: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
+/// CHECK-NOT: Callback Submit: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=1 CHECK-NOT: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
+/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3 CHECK-NOT: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
+/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4 CHECK-NOT: Callback Target:
+/// target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=2
 
 /// CHECK-NOT: Callback Target: target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=1
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
-/// CHECK-NOT: Callback Submit: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=0
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
-/// CHECK-NOT: Callback Target: target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=2
-/// CHECK-NOT: Callback Fini
+/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1 CHECK-NOT: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
+/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1 CHECK-NOT: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
+/// CHECK-NOT: Callback Submit: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=0 CHECK-NOT: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
+/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3 CHECK-NOT: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
+/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4 CHECK-NOT: Callback Target:
+/// target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=2 CHECK-NOT: Callback Fini

@doru1004
Copy link
Contributor

Libomptarget has been written for many years by many people. I would like to understand, does GH allow you to transfer over the history of contributions for each file you are moving?

@doru1004 doru1004 self-requested a review December 12, 2023 14:44
@ldionne
Copy link
Member

ldionne commented Dec 12, 2023

Libomptarget has been written for many years by many people. I would like to understand, does GH allow you to transfer over the history of contributions for each file you are moving?

Git normally handles file renaming quite well, so I would assume you won't even notice this when doing e.g. git blame.

The bigger question I'd have is about creating a new top-level directory in the monorepo -- typically this has been done mainly for new projects (e.g. llvm-libc, mlir, etc) and I wonder if there's some sort of process to add a new one of those. I don't care strongly, just mentioning it.

@doru1004
Copy link
Contributor

Git normally handles file renaming quite well, so I would assume you won't even notice this when doing e.g. git blame.

Thank you for clarifying, sounds good.

@jplehr
Copy link
Contributor

jplehr commented Dec 18, 2023

I gave this a shot with our buildbot CMake invocation to see what would happen and what we need to change, once this lands.
The CMake line the bot uses was modified to add the offload runtime, all else is the same.

cmake -G Ninja ../llvm -DCMAKE_BUILD_TYPE=Release -DLLVM_ENABLE_ASSERTIONS=ON '-DLLVM_LIT_ARGS=-vv --show-unsupported --show-xfail -j 64' -DCMAKE_INSTALL_PREFIX=/tmp/llvm.inst -DCMAKE_BUILD_TYPE=Release '-DLLVM_TARGETS_TO_BUILD=X86;AMDGPU' '-DLLVM_ENABLE_RUNTIMES=offload;openmp' '-DLLVM_ENABLE_PROJECTS=clang;lld;llvm;compiler-rt'

I get errors about not finding something

CMake Error at $HOME/git/llvm-project/offload/CMakeLists.txt:126 (include):
  include could not find requested file:

    OpenMPTesting

I also get the following that I'm not sure what to do about (I don't get this in our regular buildbot case)

CMake Warning at $HOME/git/llvm-project/offload/CMakeLists.txt:182 (message):
  LTO is not supported: Change Dir:
  $HOME/git/llvm-project/botbuild/runtimes/runtimes-bins/offload/CMakeFiles/_CMakeLTOTest-CXX/bin

  Run Build Command(s):/usr/bin/ninja && [1/4] Building CXX object
  CMakeFiles/foo.dir/foo.cpp.o

  [2/4] Building CXX object CMakeFiles/boo.dir/main.cpp.o

  [3/4] Linking CXX static library libfoo.a

  [4/4] Linking CXX executable boo

  FAILED: boo

  : && $HOME/git/llvm-project/botbuild/./bin/clang++ -flto=thin
  CMakeFiles/boo.dir/main.cpp.o -o boo libfoo.a && :

  /usr/bin/ld:
  $HOME/git/llvm-project/botbuild/bin/../lib/LLVMgold.so: error
  loading plugin:
  $HOME/git/llvm-project/botbuild/bin/../lib/LLVMgold.so: cannot
  open shared object file: No such file or directory

  clang++: error: linker command failed with exit code 1 (use -v to see
  invocation)

  ninja: build stopped: subcommand failed.

@jdoerfert
Copy link
Member Author

The first one I'll look into.
The second one is something you should see already in trunk. It's the test for LTO, though, arguably, I should not just emit the entire warning if it failed. I'll cut that down.

@jdoerfert jdoerfert force-pushed the llvm_offload branch 2 times, most recently from ebde789 to a63e186 Compare March 15, 2024 22:32
jplehr added a commit to jplehr/llvm-zorg that referenced this pull request Mar 19, 2024
This enables the new "offload" runtime that
llvm/llvm-project#75125 will create. In
addition, it adds a new lit test target.
I removed the extra CMake argument that would again specify which
runtimes to build manually, so that we rely on the factory to add the
respective runtimes.

Since the PR does not make `check-libomptarget` available in the
top level configuration, we have to change the ninja working directory
into the runtimes build folder, where the target is known. Whether or
not this works with this crude "additional lit checks" mechanism is an
experiment.
jplehr added a commit to jplehr/llvm-zorg that referenced this pull request Mar 22, 2024
This adds a new worker for AMD GPU workloads and uses that worker for
the new offload-runtime buildbot.
That builder is in preparation for landing
llvm/llvm-project#75125.

The builder enables clang and flang frontend together with the runtimes
'offload' and 'openmp' to build host-side OpenMP and GPU offloading
OpenMP.
It runs 'check-openmp' for host-side tests, 'check-clang' and
'check-flang' for frontend/codegen tests and 'check-offload'
for runtime offloading tests.
jplehr added a commit to jplehr/llvm-zorg that referenced this pull request Mar 25, 2024
This adds a new worker for AMD GPU workloads and uses that worker for
the new offload-runtime buildbot.
That builder is in preparation for landing
llvm/llvm-project#75125.

The builder enables clang and flang frontend together with the runtimes
'offload' and 'openmp' to build host-side OpenMP and GPU offloading
OpenMP.
It runs 'check-openmp' for host-side tests, 'check-clang' and
'check-flang' for frontend/codegen tests and 'check-offload'
for runtime offloading tests.
jplehr added a commit to llvm/llvm-zorg that referenced this pull request Mar 25, 2024
This adds a new worker for AMD GPU workloads and uses that worker for
the new offload-runtime buildbot.
That builder is in preparation for landing
llvm/llvm-project#75125.

The builder enables clang and flang frontend together with the runtimes
'offload' and 'openmp' to build host-side OpenMP and GPU offloading
OpenMP.
It runs 'check-openmp' for host-side tests, 'check-clang' and
'check-flang' for frontend/codegen tests and 'check-offload' for runtime
offloading tests.
@jdoerfert jdoerfert changed the title [Offload][WIP] Move /openmp/libomptarget to /offload [Offload] Move /openmp/libomptarget to /offload Mar 27, 2024
@jplehr
Copy link
Contributor

jplehr commented Apr 3, 2024

I tested this again with the buildbot config that we use and it seems that OMPT support got dropped.

In current libomptarget we have this logic https://github.com/llvm/llvm-project/blob/fa2bbea14df3273b3403f34cc295c56233fdbd0d/openmp/libomptarget/CMakeLists.txt#L104C1-L120C8 but that is not present in the new CMake file.

The result is that OMPT is not built by default, as it is today. I am not sure if we have a switch to turn it on in offload. In any case, I feel that we should aim to maintain the defaults that we have today, i.e., OMPT support on-by-default.

@saiislam
Copy link
Contributor

saiislam commented Apr 4, 2024

I tested this again with the buildbot config that we use and it seems that OMPT support got dropped.

In current libomptarget we have this logic https://github.com/llvm/llvm-project/blob/fa2bbea14df3273b3403f34cc295c56233fdbd0d/openmp/libomptarget/CMakeLists.txt#L104C1-L120C8 but that is not present in the new CMake file.

The result is that OMPT is not built by default, as it is today. I am not sure if we have a switch to turn it on in offload. In any case, I feel that we should aim to maintain the defaults that we have today, i.e., OMPT support on-by-default.

One way to solve it is to propagate the value of LIBOMP_HAVE_OMPT_SUPPORT to llvm-project/offload/CMakeLists.txt using:

set(LIBOMP_HAVE_OMPT_SUPPORT ${LIBOMP_HAVE_OMPT_SUPPORT} PARENT_SCOPE)

Above line needs to be added to openmp/CMakeLists.txt and openmp/runtime/CMakeLists.txt

PS - I am still looking into OPENMP_STANDALONE_BUILD failure.

@jdoerfert
Copy link
Member Author

jdoerfert commented Apr 6, 2024

@jdoerfert I think you additionally need the cmake changes @jplehr posted a couple of days back. This one:

In current libomptarget we have this logic https://github.com/llvm/llvm-project/blob/fa2bbea14df3273b3403f34cc295c56233fdbd0d/openmp/libomptarget/CMakeLists.txt#L104C1-L120C8 but that is not present in the new CMake file.

Without these changes, I think you will see the OMPT tests as unsupported.

I just did a build from scratch, this is what I got:

-- OMPT target enabled
llvm-lit runtimes/runtimes-bins/offload/test/amdgcn-amd-amdhsa/ompt                                                              
-- Testing: 10 tests, 10 workers --
PASS: libomptarget :: amdgcn-amd-amdhsa :: ompt/target_memcpy_emi.c (1 of 10)
PASS: libomptarget :: amdgcn-amd-amdhsa :: ompt/target_memcpy.c (2 of 10)
PASS: libomptarget :: amdgcn-amd-amdhsa :: ompt/veccopy_data.c (3 of 10)
PASS: libomptarget :: amdgcn-amd-amdhsa :: ompt/veccopy_no_device_init.c (4 of 10)
PASS: libomptarget :: amdgcn-amd-amdhsa :: ompt/veccopy_map.c (5 of 10)
PASS: libomptarget :: amdgcn-amd-amdhsa :: ompt/veccopy.c (6 of 10)
PASS: libomptarget :: amdgcn-amd-amdhsa :: ompt/veccopy_emi_map.c (7 of 10)
PASS: libomptarget :: amdgcn-amd-amdhsa :: ompt/veccopy_emi.c (8 of 10)
PASS: libomptarget :: amdgcn-amd-amdhsa :: ompt/veccopy_wrong_return.c (9 of 10)
PASS: libomptarget :: amdgcn-amd-amdhsa :: ompt/veccopy_disallow_both.c (10 of 10)

Testing Time: 3.70s

Total Discovered Tests: 10
  Passed: 10 (100.00%)

offload/CMakeLists.txt always contained:

# OMPT support for libomptarget
# Follow host OMPT support and check if host support has been requested.
# LIBOMP_HAVE_OMPT_SUPPORT indicates whether host OMPT support has been implemented.
# LIBOMP_OMPT_SUPPORT indicates whether host OMPT support has been requested (default is ON).
# LIBOMPTARGET_OMPT_SUPPORT indicates whether target OMPT support has been requested (default is ON).
set(OMPT_TARGET_DEFAULT FALSE)
if ((LIBOMP_HAVE_OMPT_SUPPORT) AND (LIBOMP_OMPT_SUPPORT) AND (NOT WIN32))
  set (OMPT_TARGET_DEFAULT TRUE)
endif()
set(LIBOMPTARGET_OMPT_SUPPORT ${OMPT_TARGET_DEFAULT} CACHE BOOL "OMPT-target-support?")
if ((OMPT_TARGET_DEFAULT) AND (LIBOMPTARGET_OMPT_SUPPORT))
  add_definitions(-DOMPT_SUPPORT=1)
  message(STATUS "OMPT target enabled")
else()
  set(LIBOMPTARGET_OMPT_SUPPORT FALSE)
  message(STATUS "OMPT target disabled")
endif()

@dhruvachak
Copy link
Contributor

Thanks @jdoerfert
Yes, I see that change now in offload/CMakeLists.txt. @jplehr Please verify you see the OMPT tests passing now in the buildbot.

saiislam added a commit to saiislam/llvm-project that referenced this pull request Apr 17, 2024
Moved some of the common functionality in a top-level common
cmake module and fixed path and inclusion issues.

Depends on llvm#75125
saiislam added a commit to saiislam/llvm-project that referenced this pull request Apr 17, 2024
The standalone offload build will just search for
the ompt header and if found then OMPT support will
be enabled. This caused the LIBOMP_HAVE_VERSION_SCRIPT_FLAG
detection and pythonize_bool to move into offload/CMakeLists.txt.

If the cmake compiler is clang, get the resource directory for
installation of openmp headers. find_package(LLVM) is already
called in LibomptargetGetDependencies.

Depends on llvm#75125
jdoerfert and others added 2 commits April 18, 2024 15:47
In a nutshell, this moves our libomptarget code to create the offload
subproject.

With this commit, users need to enable the new LLVM/Offload subproject
as a runtime in their cmake configuration. No further changes are
expected for downstream code.

Tests and other components still depend on OpenMP and have also not been
renamed. The results below are for a build in which OpenMP and Offload
are enabled runtimes. In addition to the pure git mv, I needed to adjust
some CMake files. Nothing is intended to change semantics.

```
ninja check-offload
```
Works with the X86 and AMDGPU offload tests

```
ninja check-openmp
```
Still works but doesn't build offload tests anymore.

```
ls install/lib
```
Shows all expected libraries, incl.
- `libomptarget.devicertl.a`
- `libomptarget-nvptx-sm_90.bc`
- `libomptarget.rtl.amdgpu.so` -> `libomptarget.rtl.amdgpu.so.19git`
- `libomptarget.so` -> `libomptarget.so.19git`

Fixes: llvm#75124
The standalone offload build will just search for
the ompt header and if found then OMPT support will
be enabled. This caused the LIBOMP_HAVE_VERSION_SCRIPT_FLAG
detection and pythonize_bool to move into offload/CMakeLists.txt.

If the cmake compiler is clang, get the resource directory for
installation of openmp headers. find_package(LLVM) is already
called in LibomptargetGetDependencies.

Depends on llvm#75125
Copy link
Contributor

@saiislam saiislam left a comment

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

LGTM.

Thanks for the support and patience!

Copy link
Contributor

@jplehr jplehr left a comment

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

I think we are good from the build bots point of view.

@jdoerfert jdoerfert merged commit 330d898 into llvm:main Apr 22, 2024
@ye-luo
Copy link
Contributor

ye-luo commented Apr 22, 2024

libomptarget files install location is changed back to <install>/lib from <install>/lib/x86_64-unknown-linux-gnu

@jhuber6
Copy link
Contributor

jhuber6 commented Apr 22, 2024

libomptarget files install location is changed back to <install>/lib from <install>/lib/x86_64-unknown-linux-gnu

I'm waiting for someone to review #89645

searlmc1 pushed a commit to ROCm/llvm-project that referenced this pull request Apr 23, 2024
Based on PR llvm#75125

In a nutshell, this moves our libomptarget code to create the offload
subproject. For now, we allow LLVM/Offload only as runtime. Tests and
other components still depend on OpenMP and have also not been
renamed. The results below are for a build in which OpenMP and
Offload are enabled runtimes. In addition to the pure git mv, I
needed to adjust some CMake files. Nothing is intended to change
semantics but some of the things likely broke other build
configurations. Testers are needed.

ninja check-offload
Works with the X86 and AMDGPU offload tests

ninja check-openmp
Still works but doesn't build offload tests anymore.

ls install/lib
Shows all expected libraries, incl.

libomptarget.devicertl.a
libomptarget-nvptx-sm_90.bc
libomptarget.rtl.amdgpu.so -> libomptarget.rtl.amdgpu.so.18git
libomptarget.so -> libomptarget.so.18git

Fixes: llvm#75124
Change-Id: Ie1d7a9805309c63292e6378f077ad319b42dd279
RESULT_VARIABLE COMMAND_RETURN_CODE
OUTPUT_VARIABLE COMPILER_RESOURCE_DIR
)
set(LIBOMP_HEADERS_INSTALL_PATH "${COMPILER_RESOURCE_DIR}/include")
Copy link
Member

@daltenty daltenty Apr 29, 2024

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

The addition of this change seems to have broken the stand alone build we were doing:

-- Installing: /nfs/build//lib/clang/19/include/omp.h
CMake Error at runtime/src/cmake_install.cmake:101 (file):
  file INSTALL cannot copy file
  "/home/daltenty/build-openmp/32/runtime/src/omp.h" to
  "/nfs/build/lib/clang/19/include/omp.h":
  Read-only file system.
Call Stack (most recent call first):
  runtime/cmake_install.cmake:42 (include)
  cmake_install.cmake:42 (include)

We're now trying to install to the build compiler's resource directory for some reason, which is read-only and isn't really what we intended here.

Copy link
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Yeah that's definitely wrong for a standalone build. For a runtimes build it's so we install it in a place where the compiler can find it so ./bin/clang actually works. We'll need to fix a lot of this.

Copy link
Member

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

I think there's some confusion here about the effects of this.

Just for terminology clarity, my understanding is there are roughly three ways we can build runtimes subprojects of LLVM (defined by the top-level CMake source directory and CMake options you use):

  1. Projects+Runtimes build with Clang (i.e. ):
    cmake -S llvm-project/llvm -DENABLE_PROJECTS=clang -DENABLE_RUNTIMES=openmp
  2. Runtimes-only build
    cmake -S llvm-project/runtimes -DENABLE_RUNTIMES=openmp
  3. Stand-alone build (from before the monorepo existed, doesn't really work for all subprojects anymore)
    cmake -S llvm-project/openmp -DOPENMP_LLVM_TOOLS_DIR=/path/to/llvm/build/bin

Yeah that's definitely wrong for a standalone build. For a runtimes build it's so we install it in a place where the compiler can find it so ./bin/clang actually works. We'll need to fix a lot of this.

I assume this code was intended for a Projects+Runtimes build with clang. But it's inside a if(OPENMP_STANDALONE_BUILD) block which only seems to be true for standalone builds.

There are some misleading comments in the source about this here

# llvm/runtimes/ will set OPENMP_STANDALONE_BUILD.
but I checked my builds and I don't see it being set by llvm/runtimes at all and seems to be false during the tests. So I'm not sure this doing what was intended.

Copy link
Member

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Posted #91243 to address this

@fhahn
Copy link
Contributor

fhahn commented Apr 30, 2024

With this change, previous cmake configurations only using -DLLVM_ENABLE_RUNTIMES="openmp" silently appear to not configure openmp at all; if offload is required for openmp to be enabled, could we error on -DLLVM_ENABLE_RUNTIMES="openmp"?

It also looks like https://openmp.llvm.org/SupportAndFAQ.html needs updating?

@jdoerfert
Copy link
Member Author

@jhuber6 @jplehr Are you already looking at the reported issues above?

daltenty added a commit to daltenty/llvm-project that referenced this pull request May 6, 2024
…hanges

Revert the portion of llvm#75125 which touch the LIBOMP_HEADERS_INSTALL_PATH in standalone build.

This change is harmful for standalone builds, since it tries to overwrite the omp.h inside the build compiler. For all-in-one builds, I've confirmed this change is unncessary as llvm#88007 already set it up so that omp.h will be put into the project build clang's resource directory.
daltenty added a commit that referenced this pull request May 6, 2024
…91243)

Revert the portion of #75125
which modified the LIBOMP_HEADERS_INSTALL_PATH in standalone build.

This change is harmful for real standalone builds (i.e. builds where we
build openmp by itself), since it tries to overwrite the `omp.h` inside
the build compiler. For all-in-one builds with clang, testing shows this
change is unnecessary as #88007
already set up that build configuration so that omp.h will be put into
the project build's `clang` resource directory.
jprotze added a commit that referenced this pull request Jun 6, 2024
PR #75125 introduced upward propagation of some OMPT-related CMake
variables.
For stand-alone builds this results in a warning that `SCOPE_PARENT` has
no meaning in a top-level directory.
Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment
Labels
backend:AMDGPU cmake Build system in general and CMake in particular offload openmp:libomp OpenMP host runtime openmp:libomptarget OpenMP offload runtime
Projects
Status: Done
Development

Successfully merging this pull request may close these issues.

[Offload] Create a PR to move /openmp/libomptarget to /offload