Skip to content
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

Add common interface for simple kernels #733

Merged
merged 56 commits into from Jul 10, 2021
Merged
Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
56 commits
Select commit Hold shift + click to select a range
2d7f0ea
add simple kernel support
upsj Mar 24, 2021
7392892
Revert "add simple kernel support"
upsj Apr 2, 2021
e65792d
add direct kernel launch wrappers
upsj Apr 2, 2021
3e08abf
set correct nvcc flags
upsj Apr 2, 2021
1e8fae1
work around icpc internal compiler error
upsj Apr 2, 2021
4ea1bf0
workaround for MSVC extended-lambda issue
upsj Apr 2, 2021
9785d51
unify cg and precision_conversion kernel files
upsj May 21, 2021
3b878b1
review updates and formatting
upsj May 21, 2021
de7a8f4
unify kernel files
upsj May 21, 2021
225a78d
make simple kernel launchers free functions
upsj May 21, 2021
dfe928e
make device mapping more flexible
upsj May 21, 2021
c46e94d
move dense operations to simple interface
upsj May 21, 2021
3359b75
add simple Krylov kernels
upsj May 21, 2021
fff356c
eliminate unnecessary stride parameters
upsj May 22, 2021
4863109
replace BLAS axpy and scal by simple kernels
upsj May 22, 2021
c84c707
don't preserve stride on create_with_same_config
upsj Jun 1, 2021
1bb3100
review updates
upsj Jun 1, 2021
147745f
unify simple solver kernel tests
upsj Jun 8, 2021
d8724ed
add tests for stride handling and zero division
upsj Jun 8, 2021
cd8143b
review updates
upsj Jun 8, 2021
b4225bf
use consistent zero assignment in common kernels
upsj Jun 8, 2021
1ac5aa8
remove unstable bicg test
upsj Jun 8, 2021
33254fb
add kernel launch tests
upsj Jun 9, 2021
415c41b
simplify device_unpack_2d_impl parameters
upsj Jun 11, 2021
6e8ec19
adapt test_install paths
upsj Jun 11, 2021
cf11409
fix kernel launch tests
upsj Jun 11, 2021
4863501
add support for DPCPP_SINGLE_MODE to tests
upsj Jun 11, 2021
e2dd2aa
fix clang-cuda tests
upsj Jun 11, 2021
a1e93c6
fix OpenMP collapse compilation issue
upsj Jun 11, 2021
6de0b69
relax solver test precision
upsj Jun 11, 2021
1e9f716
use SYCL_DEVICE_TYPE/FILTER to make sure using gpu/cpu
yhmtsai May 20, 2021
f6c6d98
remove OpenMP collapse(2)
upsj Jun 11, 2021
70c16d7
explicitly unroll OpenMP simple kernels
upsj Jun 11, 2021
e4879b6
avoid unrolling remainder loop
upsj Jun 11, 2021
39b7ed3
unroll width 4
upsj Jun 11, 2021
64cccd7
explicitly unroll remainder with width 4
upsj Jun 11, 2021
000b973
add kernel_launch to format_header.sh config
upsj Jun 11, 2021
ef37edb
keep stride in solver create_with_same_config
upsj Jun 17, 2021
2aa080f
update documentation and assertions
upsj Jun 17, 2021
3630a7a
fix remaining create_with_same_config test
upsj Jun 17, 2021
1c87a35
review updates
upsj Jun 23, 2021
87287a4
extract common code for simple kernel launch
upsj Jun 23, 2021
a4d6278
work around DPC++ use-after-free
upsj Jun 23, 2021
1d29893
improve documentation for kernel_launch
upsj Jun 24, 2021
1cb4705
move simple dense kernel tests to common tests
upsj Jun 24, 2021
9c5df14
increase BiCGSTAB test tolerances
upsj Jun 24, 2021
5dc8aac
fix formatting
upsj Jun 24, 2021
8e3a2bb
move kernel_launch defines into test CMakeLists
upsj Jun 24, 2021
56948b3
update test_install path for formatting job
upsj Jun 29, 2021
dc6f4d7
rename compile tests inside test/
upsj Jun 29, 2021
382f501
remove circular dependency in kernel_launch header
upsj Jun 29, 2021
c29627e
review updates
upsj Jun 29, 2021
0ef51f1
remove magic numbers for OpenMP unrolling
upsj Jun 29, 2021
65a0a65
fix include position for kernel_launch tests
upsj Jun 29, 2021
8828510
fix single-precision dpcpp dense kernel tests
upsj Jun 29, 2021
29af274
skip DPCPP solver apply tests
upsj Jul 9, 2021
File filter

Filter by extension

Filter by extension

Conversations
Failed to load comments.
Jump to
Jump to file
Failed to load files.
Diff view
Diff view
2 changes: 1 addition & 1 deletion .github/bot-pr-format-base.sh
Expand Up @@ -4,7 +4,7 @@ source .github/bot-pr-base.sh

EXTENSION_REGEX='\.(cuh?|hpp|hpp\.inc?|cpp)$'
FORMAT_HEADER_REGEX='^(benchmark|core|cuda|hip|include/ginkgo/core|omp|reference|dpcpp)/'
FORMAT_REGEX='^(common|examples|test_install)/'
FORMAT_REGEX='^(common|examples|test)/'

echo "Retrieving PR file list"
PR_FILES=$(bot_get_all_changed_files ${PR_URL})
Expand Down
8 changes: 5 additions & 3 deletions .gitlab-ci.yml
Expand Up @@ -98,6 +98,7 @@ include:
CUDA_ARCH_STR=-DGINKGO_CUDA_ARCHITECTURES=${CUDA_ARCH};
CUDA_HOST_STR=-DCMAKE_CUDA_HOST_COMPILER=$(which ${CXX_COMPILER});
fi
- if [ -n "${SYCL_DEVICE_TYPE}" ]; then export SYCL_DEVICE_TYPE; fi
- if [ -n "${SYCL_DEVICE_FILTER}" ]; then export SYCL_DEVICE_FILTER; fi
- cmake ${CI_PROJECT_DIR}${CI_PROJECT_DIR_SUFFIX}
-GNinja
Expand All @@ -121,7 +122,7 @@ include:
(( $(ctest -N | tail -1 | sed 's/Total Tests: //') != 0 )) || exit 1
- ctest -V
- ninja test_install
- pushd test_install
- pushd test/test_install
- ninja install
- popd
- |
Expand All @@ -145,6 +146,7 @@ include:
ninja validate_all_examples
fi
fi
- if [ -n "${SYCL_DEVICE_TYPE}" ]; then unset SYCL_DEVICE_TYPE; fi
- if [ -n "${SYCL_DEVICE_FILTER}" ]; then unset SYCL_DEVICE_FILTER; fi
- if [ "${EXPORT_BUILD_DIR}" == "ON" ]; then ninja test_exportbuild; fi
dependencies: []
Expand Down Expand Up @@ -718,7 +720,7 @@ build/dpcpp/cpu/release/static:
BUILD_DPCPP: "ON"
BUILD_TYPE: "Release"
BUILD_SHARED_LIBS: "ON"
SYCL_DEVICE_FILTER: "CPU"
SYCL_DEVICE_TYPE: "CPU"

build/dpcpp/igpu/release/static:
<<: *default_build_with_test
Expand All @@ -733,7 +735,7 @@ build/dpcpp/igpu/release/static:
BUILD_TYPE: "Release"
BUILD_SHARED_LIBS: "OFF"
DPCPP_SINGLE_MODE: "ON"
SYCL_DEVICE_FILTER: "GPU"
SYCL_DEVICE_FILTER: "Level_Zero:GPU"

build/dpcpp/dgpu/debug/shared:
<<: *default_build_with_test
Expand Down
66 changes: 41 additions & 25 deletions CMakeLists.txt
Expand Up @@ -261,6 +261,9 @@ if (GINKGO_BUILD_OMP)
endif()
add_subdirectory(core) # Core Ginkgo types and top-level functions
add_subdirectory(include) # Public API self-contained check
if (GINKGO_BUILD_TESTS)
add_subdirectory(test) # Tests running on all executors
endif()

# Non core directories and targets
if(GINKGO_BUILD_EXAMPLES)
Expand Down Expand Up @@ -330,56 +333,69 @@ ginkgo_modify_flags(CMAKE_CUDA_FLAGS_DEBUG)
ginkgo_modify_flags(CMAKE_CUDA_FLAGS_RELEASE)
ginkgo_install()

set(GINKGO_TEST_INSTALL_SRC_DIR "${Ginkgo_SOURCE_DIR}/test/test_install/")
set(GINKGO_TEST_INSTALL_BIN_DIR "${Ginkgo_BINARY_DIR}/test/test_install/")
set(GINKGO_TEST_EXPORTBUILD_SRC_DIR "${Ginkgo_SOURCE_DIR}/test/test_exportbuild/")
set(GINKGO_TEST_EXPORTBUILD_BIN_DIR "${Ginkgo_BINARY_DIR}/test/test_exportbuild/")
if(MSVC)
# Set path/command with $<CONFIG>
set(GINKGO_TEST_INSTALL_COMMAND "${Ginkgo_BINARY_DIR}/test_install/$<CONFIG>/test_install")
set(GINKGO_TEST_EXPORTBUILD_COMMAND "${Ginkgo_BINARY_DIR}/test_exportbuild/$<CONFIG>/test_exportbuild")
set(GINKGO_TEST_INSTALL_CMD ${GINKGO_TEST_INSTALL_BIN_DIR}/$<CONFIG>/test_install)
set(GINKGO_TEST_EXPORTBUILD_CMD ${GINKGO_TEST_EXPORTBUILD_BIN_DIR}/$<CONFIG>/test_exportbuild)
thoasm marked this conversation as resolved.
Show resolved Hide resolved
if(GINKGO_BUILD_CUDA)
set(GINKGO_TEST_INSTALL_CUDA_COMMAND "${Ginkgo_BINARY_DIR}/test_install/$<CONFIG>/test_install_cuda")
set(GINKGO_TEST_INSTALL_CUDA_CMD ${GINKGO_TEST_INSTALL_BIN_DIR}/$<CONFIG>/test_install_cuda)
endif()
else()
set(GINKGO_TEST_INSTALL_COMMAND "${Ginkgo_BINARY_DIR}/test_install/test_install")
set(GINKGO_TEST_EXPORTBUILD_COMMAND "${Ginkgo_BINARY_DIR}/test_exportbuild/test_exportbuild")
set(GINKGO_TEST_INSTALL_CMD ${GINKGO_TEST_INSTALL_BIN_DIR}/test_install)
set(GINKGO_TEST_EXPORTBUILD_CMD ${GINKGO_TEST_EXPORTBUILD_BIN_DIR}/test_exportbuild)
if(GINKGO_BUILD_CUDA)
set(GINKGO_TEST_INSTALL_CUDA_COMMAND "${Ginkgo_BINARY_DIR}/test_install/test_install_cuda")
endif()
if(GINKGO_BUILD_HIP)
set(GINKGO_TEST_INSTALL_HIP_COMMAND "${Ginkgo_BINARY_DIR}/test_install/test_install_hip")
set(GINKGO_TEST_INSTALL_CUDA_CMD ${GINKGO_TEST_INSTALL_BIN_DIR}/test_install_cuda)
endif()
endif()
if(GINKGO_BUILD_HIP)
set(GINKGO_TEST_INSTALL_HIP_CMD ${GINKGO_TEST_INSTALL_BIN_DIR}/test_install_hip)
endif()

file(MAKE_DIRECTORY ${Ginkgo_BINARY_DIR}/test_install)
file(MAKE_DIRECTORY "${GINKGO_TEST_INSTALL_BIN_DIR}")
file(MAKE_DIRECTORY "${GINKGO_TEST_EXPORTBUILD_BIN_DIR}")
set(TOOLSET "")
if (NOT "${CMAKE_GENERATOR_TOOLSET}" STREQUAL "")
set(TOOLSET "-T${CMAKE_GENERATOR_TOOLSET}")
endif()
add_custom_target(test_install
COMMAND ${CMAKE_COMMAND} -G${CMAKE_GENERATOR} ${TOOLSET} -H${Ginkgo_SOURCE_DIR}/test_install
-B${Ginkgo_BINARY_DIR}/test_install
COMMAND ${CMAKE_COMMAND} -G${CMAKE_GENERATOR} ${TOOLSET}
-H${GINKGO_TEST_INSTALL_SRC_DIR}
-B${GINKGO_TEST_INSTALL_BIN_DIR}
-DCMAKE_BUILD_TYPE=${CMAKE_BUILD_TYPE}
-DCMAKE_PREFIX_PATH=${CMAKE_INSTALL_PREFIX}/${GINKGO_INSTALL_CONFIG_DIR}
-DCMAKE_CXX_COMPILER=${CMAKE_CXX_COMPILER} -DCMAKE_C_COMPILER=${CMAKE_C_COMPILER}
-DCMAKE_CXX_COMPILER=${CMAKE_CXX_COMPILER}
-DCMAKE_C_COMPILER=${CMAKE_C_COMPILER}
-DCMAKE_CUDA_COMPILER=${CMAKE_CUDA_COMPILER}
WORKING_DIRECTORY ${Ginkgo_BINARY_DIR}/test_install
# `--config cfg` is ignored by single-configuration generator.
# `$<CONFIG>` is always be the same as `CMAKE_BUILD_TYPE` in
# single-configuration generator.
COMMAND ${CMAKE_COMMAND} --build ${Ginkgo_BINARY_DIR}/test_install --config $<CONFIG>
COMMAND ${GINKGO_TEST_INSTALL_COMMAND}
COMMAND ${GINKGO_TEST_INSTALL_CUDA_COMMAND}
COMMAND ${GINKGO_TEST_INSTALL_HIP_COMMAND}
COMMENT "Running a test on the installed binaries. This requires running `(sudo) make install` first.")
COMMAND ${CMAKE_COMMAND}
--build ${GINKGO_TEST_INSTALL_BIN_DIR}
--config $<CONFIG>
COMMAND ${GINKGO_TEST_INSTALL_CMD}
COMMAND ${GINKGO_TEST_INSTALL_CUDA_CMD}
COMMAND ${GINKGO_TEST_INSTALL_HIP_CMD}
WORKING_DIRECTORY ${GINKGO_TEST_INSTALL_BIN_DIR}
COMMENT "Running a test on the installed binaries. "
"This requires running `(sudo) make install` first.")

add_custom_target(test_exportbuild
COMMAND ${CMAKE_COMMAND} -G${CMAKE_GENERATOR} ${TOOLSET} -H${Ginkgo_SOURCE_DIR}/test_exportbuild
-B${Ginkgo_BINARY_DIR}/test_exportbuild
-DCMAKE_CXX_COMPILER=${CMAKE_CXX_COMPILER} -DCMAKE_C_COMPILER=${CMAKE_C_COMPILER}
COMMAND ${CMAKE_COMMAND} -G${CMAKE_GENERATOR} ${TOOLSET}
-H${GINKGO_TEST_EXPORTBUILD_SRC_DIR}
-B${GINKGO_TEST_EXPORTBUILD_BIN_DIR}
-DCMAKE_CXX_COMPILER=${CMAKE_CXX_COMPILER}
-DCMAKE_C_COMPILER=${CMAKE_C_COMPILER}
-DCMAKE_CUDA_COMPILER=${CMAKE_CUDA_COMPILER}
# `--config cfg` is ignored by single-configuration generator.
# `$<CONFIG>` is always be the same as `CMAKE_BUILD_TYPE` in
# single-configuration generator.
COMMAND ${CMAKE_COMMAND} --build ${Ginkgo_BINARY_DIR}/test_exportbuild --config $<CONFIG>
COMMAND ${GINKGO_TEST_EXPORTBUILD_COMMAND}
COMMAND ${CMAKE_COMMAND}
--build ${GINKGO_TEST_EXPORTBUILD_BIN_DIR}
--config $<CONFIG>
COMMAND ${GINKGO_TEST_EXPORTBUILD_CMD}
COMMENT "Running a test on Ginkgo's exported build directory. "
"This requires compiling Ginkgo with `-DGINKGO_EXPORT_BUILD_DIR=ON` first.")

Expand Down
14 changes: 10 additions & 4 deletions cmake/build_helpers.cmake
Expand Up @@ -43,20 +43,23 @@ function(ginkgo_compile_features name)
set_target_properties("${name}" PROPERTIES POSITION_INDEPENDENT_CODE ON)
endfunction()

function(ginkgo_check_headers target)
function(ginkgo_check_headers target defines)
# build object library used to "compile" the headers
# add a proxy source file for each header in the target source list
file(GLOB_RECURSE CUDA_HEADERS RELATIVE "${CMAKE_CURRENT_SOURCE_DIR}" CONFIGURE_DEPENDS "*.cuh")
file(GLOB_RECURSE HIP_HEADERS RELATIVE "${CMAKE_CURRENT_SOURCE_DIR}" CONFIGURE_DEPENDS "*.hip.hpp")
file(GLOB_RECURSE CXX_HEADERS RELATIVE "${CMAKE_CURRENT_SOURCE_DIR}" CONFIGURE_DEPENDS "*.hpp")
list(FILTER CXX_HEADERS EXCLUDE REGEX ".*\.hip\.hpp$")
list(FILTER CXX_HEADERS EXCLUDE REGEX "^test.*")
list(FILTER CXX_HEADERS EXCLUDE REGEX "^base/kernel_launch.*")
list(FILTER CUDA_HEADERS EXCLUDE REGEX "^test.*")
list(FILTER CUDA_HEADERS EXCLUDE REGEX "^base/kernel_launch.*")
list(FILTER HIP_HEADERS EXCLUDE REGEX "^test.*")
list(FILTER HIP_HEADERS EXCLUDE REGEX "^base/kernel_launch.*")

set(SOURCES "")
# if we have any CUDA files in there, compile everything as CUDA
if (CUDA_HEADERS)
if(CUDA_HEADERS)
set(CUDA_HEADERS ${CUDA_HEADERS} ${CXX_HEADERS})
set(CXX_HEADERS "")
if (HIP_HEADERS)
Expand All @@ -74,10 +77,13 @@ function(ginkgo_check_headers target)
file(WRITE "${HEADER_SOURCEFILE}" "#include \"${HEADER}\"")
list(APPEND SOURCES "${HEADER_SOURCEFILE}")
endforeach()
if (SOURCES)
if(SOURCES)
add_library(${target}_headers OBJECT ${SOURCES})
target_link_libraries(${target}_headers PRIVATE ${target})
target_include_directories(${target}_headers PRIVATE "${CMAKE_CURRENT_SOURCE_DIR}")
if(defines)
target_compile_definitions(${target}_headers PRIVATE ${defines})
endif()
endif()

set(HIP_SOURCES "")
Expand All @@ -86,7 +92,7 @@ function(ginkgo_check_headers target)
file(WRITE "${HEADER_SOURCEFILE}" "#include \"${HEADER}\"")
list(APPEND HIP_SOURCES "${HEADER_SOURCEFILE}")
endforeach()
if (HIP_SOURCES)
if(HIP_SOURCES)
set_source_files_properties(${HIP_SOURCES} PROPERTIES HIP_SOURCE_PROPERTY_FORMAT TRUE)
hip_add_library(${target}_headers_hip ${HIP_SOURCES}) # the compiler options get set by linking to ginkgo_hip
target_link_libraries(${target}_headers_hip PRIVATE ${target} roc::hipblas roc::hipsparse hip::hiprand roc::rocrand)
Expand Down