diff --git a/.github/workflows/comment-bot.yml b/.github/workflows/comment-bot.yml index 6d491db..a19ef5b 100644 --- a/.github/workflows/comment-bot.yml +++ b/.github/workflows/comment-bot.yml @@ -31,6 +31,7 @@ jobs: post({ owner: context.repo.owner, repo: context.repo.repo, comment_id: context.payload.comment.id, content: "eyes"}) + github-token: ${{ secrets.GH_TOKEN || github.token }} - name: Tag Commit run: | git clone https://${GITHUB_TOKEN}@github.com/${GITHUB_REPOSITORY} repo @@ -39,7 +40,7 @@ jobs: rm -rf repo env: BODY: ${{ github.event.comment.body }} - GITHUB_TOKEN: ${{ secrets.GH_TOKEN }} + GITHUB_TOKEN: ${{ secrets.GH_TOKEN || github.token }} - name: React Success uses: actions/github-script@v7 with: @@ -50,3 +51,4 @@ jobs: post({ owner: context.repo.owner, repo: context.repo.repo, comment_id: context.payload.comment.id, content: "rocket"}) + github-token: ${{ secrets.GH_TOKEN || github.token }} diff --git a/.github/workflows/test.yml b/.github/workflows/test.yml index 6c3f1c8..a1d8b33 100644 --- a/.github/workflows/test.yml +++ b/.github/workflows/test.yml @@ -3,42 +3,17 @@ on: push: pull_request: schedule: [{cron: '30 23 * * 6'}] # M H d m w (Sat at 23:30) - workflow_dispatch: jobs: - check: + clang-format: if: github.event_name != 'pull_request' || !contains('OWNER,MEMBER,COLLABORATOR', github.event.pull_request.author_association) runs-on: ubuntu-latest steps: - uses: actions/checkout@v4 - uses: actions/setup-python@v5 - with: - python-version: '3.x' - - name: set PYSHA - run: echo "PYSHA=$(python -VV | sha256sum | cut -d' ' -f1)" >> $GITHUB_ENV - - uses: actions/cache@v3 - with: - path: ~/.cache/pre-commit - key: pre-commit|${{ env.PYSHA }}|${{ hashFiles('.pre-commit-config.yaml') }} - - name: dependencies - run: | - pip install -U pre-commit - sudo apt-get install -yqq clang-format - - uses: reviewdog/action-setup@v1 - - if: github.event_name == 'push' || github.event_name == 'pull_request' - name: comment - run: | - if [[ $EVENT == pull_request ]]; then - REPORTER=github-pr-review - else - REPORTER=github-check - fi - pre-commit run -a todo | reviewdog -efm="%f:%l: %m" -name=TODO -tee -reporter=$REPORTER -filter-mode nofilter - pre-commit run -a flake8 | reviewdog -f=pep8 -name=flake8 -tee -reporter=$REPORTER -filter-mode nofilter - pre-commit run -a mypy | reviewdog -efm="%f:%l: %m" -name=mypy -tee -reporter=$REPORTER -filter-mode nofilter - env: - REVIEWDOG_GITHUB_API_TOKEN: ${{ secrets.GITHUB_TOKEN }} - EVENT: ${{ github.event_name }} - - run: pre-commit run -a --show-diff-on-failure + with: {python-version: '3.x'} + - run: sudo apt-get install -yqq clang-format + - uses: pre-commit/action@v3.0.1 + with: {extra_args: --all-files clang-format} test: if: github.event_name != 'pull_request' || !contains('OWNER,MEMBER,COLLABORATOR', github.event.pull_request.author_association) name: py${{ matrix.python }} @@ -48,31 +23,18 @@ jobs: python: [3.7, 3.11] steps: - uses: actions/checkout@v4 - with: - fetch-depth: 0 + with: {fetch-depth: 0} - uses: actions/setup-python@v5 with: python-version: ${{ matrix.python }} - name: pip install -U -e .[dev] - # in-place for pytest (-e . doesn't work yet for scikit-build-core) - run: | - pip install toml - python -c 'import toml; c=toml.load("pyproject.toml") - print("\0".join(c["build-system"]["requires"] + ["cmake>=" + c["tool"]["scikit-build"]["cmake"]["minimum-version"]]), end="")' \ - | xargs -0 pip install ninja - pip install --no-build-isolation --no-deps -t . -U -v . -Ccmake.define.CUVEC_DEBUG=1 - git restore cuvec/src - python -c 'import toml; c=toml.load("pyproject.toml") - print("\0".join(c["project"]["dependencies"] + c["project"]["optional-dependencies"]["dev"]), end="")' \ - | xargs -0 pip install - - run: pytest - - uses: codecov/codecov-action@v3 + run: make CXX_FLAGS='' deps-build build-editable deps-run + - run: make test + - uses: codecov/codecov-action@v4 + with: + token: ${{ secrets.CODECOV_TOKEN }} - name: compile -Wall - run: | - git clean -Xdf - pip install build - python -m build -n -w \ - -Ccmake.define.CMAKE_CXX_FLAGS="-Wall -Wextra -Wpedantic -Werror -Wno-missing-field-initializers -Wno-unused-parameter -Wno-cast-function-type" + run: make clean build-wheel cuda: if: github.event_name != 'pull_request' || !contains('OWNER,MEMBER,COLLABORATOR', github.event.pull_request.author_association) name: CUDA py${{ matrix.python }} @@ -82,41 +44,25 @@ jobs: python: [3.7, 3.11] steps: - uses: actions/checkout@v4 - with: - fetch-depth: 0 + with: {fetch-depth: 0} - name: Run setup-python - run: setup-python -p${{ matrix.python }} 'cuda-version<12' cupy + run: setup-python -p${{ matrix.python }} 'cuda-version<12' 'cupy<13' - name: pip install -U -e .[dev] - # in-place for pytest (-e . doesn't work yet for scikit-build-core) - run: | - pip install toml - python -c 'import toml; c=toml.load("pyproject.toml") - print("\0".join(c["build-system"]["requires"] + ["cmake>=" + c["tool"]["scikit-build"]["cmake"]["minimum-version"]]), end="")' \ - | xargs -0 pip install ninja - pip install --no-build-isolation --no-deps -t . -U -v . -Ccmake.define.CUVEC_DEBUG=1 - git restore cuvec/src - python -c 'import toml; c=toml.load("pyproject.toml") - print("\0".join(c["project"]["dependencies"] + c["project"]["optional-dependencies"]["dev"]), end="")' \ - | xargs -0 pip install - - run: pytest - - uses: codecov/codecov-action@v3 + run: make CXX_FLAGS='' deps-build build-editable deps-run + - run: make test + - uses: codecov/codecov-action@v4 + with: + token: ${{ secrets.CODECOV_TOKEN }} - name: compile -Wall - run: | - git clean -Xdf - pip install build - python -m build -n -w \ - -Ccmake.define.CMAKE_CXX_FLAGS="-Wall -Wextra -Wpedantic -Werror -Wno-missing-field-initializers -Wno-unused-parameter -Wno-cast-function-type" \ - -Ccmake.define.CMAKE_CUDA_ARCHITECTURES=all + run: make CUDA_ARCHITECTURES=all clean build-wheel - name: Post Run setup-python run: setup-python -p${{ matrix.python }} -Dr if: ${{ always() }} deploy: - needs: [check, test, cuda] + needs: [clang-format, test, cuda] name: PyPI Deploy environment: pypi - permissions: - contents: write - id-token: write + permissions: {contents: write, id-token: write} runs-on: ubuntu-latest steps: - uses: actions/checkout@v4 @@ -124,15 +70,12 @@ jobs: fetch-depth: 0 token: ${{ secrets.GH_TOKEN || github.token }} - uses: actions/setup-python@v5 - with: - python-version: '3.x' + with: {python-version: '3.x'} - id: dist uses: casperdcl/deploy-pypi@v2 with: build: -s - upload: false - - if: github.event_name == 'push' && startsWith(github.ref, 'refs/tags') - uses: pypa/gh-action-pypi-publish@release/v1 + upload: ${{ github.event_name == 'push' && startsWith(github.ref, 'refs/tags') }} - if: github.event_name == 'push' && startsWith(github.ref, 'refs/tags') name: Release run: | @@ -142,11 +85,7 @@ jobs: env: GH_TOKEN: ${{ secrets.GH_TOKEN || github.token }} - name: Docs - run: | - pushd docs - pip install -U -r requirements.txt - PYTHONPATH=. pydoc-markdown --build --site-dir=../../../dist/site - popd + run: make deps-docs docs - if: ${{ github.event_name == 'push' && startsWith(github.ref, 'refs/tags') || github.event_name == 'workflow_dispatch' }} uses: casperdcl/push-dir@v1 with: diff --git a/.gitignore b/.gitignore index 9f801ee..4a3acf9 100644 --- a/.gitignore +++ b/.gitignore @@ -10,6 +10,7 @@ __pycache__/ /cuvec/example_swig.py /cuvec/swvec.py /dist/ +/build/ /docs/build/ # Unit test / coverage reports diff --git a/.pre-commit-config.yaml b/.pre-commit-config.yaml index 06994de..5ba422f 100644 --- a/.pre-commit-config.yaml +++ b/.pre-commit-config.yaml @@ -1,3 +1,4 @@ +ci: {skip: [clang-format]} default_language_version: python: python3 repos: @@ -26,7 +27,7 @@ repos: types: [text] exclude: ^(.pre-commit-config.yaml|.github/workflows/test.yml)$ - repo: https://github.com/PyCQA/flake8 - rev: 6.1.0 + rev: 7.0.0 hooks: - id: flake8 args: [-j8] @@ -50,7 +51,7 @@ repos: args: [-i] additional_dependencies: [toml] - repo: https://github.com/PyCQA/isort - rev: 5.12.0 + rev: 5.13.2 hooks: - id: isort - repo: https://github.com/doublify/pre-commit-clang-format diff --git a/CONTRIBUTING.md b/CONTRIBUTING.md index 058c30d..447c3aa 100644 --- a/CONTRIBUTING.md +++ b/CONTRIBUTING.md @@ -3,31 +3,24 @@ Install in "development/editable" mode including dev/test dependencies: ```sh -git clone https://github.com/AMYPAD/CuVec && cd CuVec +# clone & install dependencies (one-off) +git clone https://github.com/AMYPAD/CuVec +cd CuVec +make deps-build deps-run -# `pip install -e .[dev]` won't work due to https://github.com/scikit-build/scikit-build-core/issues/114 -# work-around: -# 1. install dependencies (one-off) -pip install toml -python -c 'import toml; c=toml.load("pyproject.toml") -print("\0".join(c["build-system"]["requires"] + c["project"]["dependencies"] + c["project"]["optional-dependencies"]["dev"]), end="")' \ -| xargs -0 pip install -U ninja cmake -# 2. delete build artefacts, (re)build & install in-place with debug info -git clean -Xdf -pip install --no-build-isolation --no-deps -t . -U -v . \ - -Ccmake.define.CUVEC_DEBUG=1 - -Ccmake.define.CMAKE_CXX_FLAGS="-Wall -Wextra -Wpedantic -Werror -Wno-missing-field-initializers -Wno-unused-parameter -Wno-cast-function-type" -git restore cuvec/src # undo deletion of sources +# delete build artefacts, (re)build & install in-place with debug info +make CUVEC_DEBUG=1 build-editable ``` Once installed in development/editable mode, tests may be run using: ```sh -pytest +pytest -k "not perf" ``` To run performance tests, build with debugging disabled (`CUVEC_DEBUG=0`), then run: ```sh +pytest -k "perf" -n=0 python tests/test_perf.py ``` diff --git a/Makefile b/Makefile new file mode 100644 index 0000000..f3abdbb --- /dev/null +++ b/Makefile @@ -0,0 +1,36 @@ +# NOTE: cannot `pip install -U -e .[dev]` (install in-place for pytest) +# since `-e .` doesn't work yet (https://github.com/scikit-build/scikit-build-core/issues/114). +# Instead, do `make deps-build build-editable deps-run` +CUVEC_DEBUG=0 +CXX_FLAGS=-Wall -Wextra -Wpedantic -Werror -Wno-missing-field-initializers -Wno-unused-parameter -Wno-cast-function-type +CUDA_ARCHITECTURES=native +BUILD_CMAKE_FLAGS=-Ccmake.define.CUVEC_DEBUG=$(CUVEC_DEBUG) -Ccmake.define.CMAKE_CXX_FLAGS="$(CXX_FLAGS)" -Ccmake.define.CMAKE_CUDA_ARCHITECTURES=$(CUDA_ARCHITECTURES) +CCACHE= +ifneq ($(CCACHE),) + BUILD_CMAKE_FLAGS+= -Ccmake.define.CMAKE_CXX_COMPILER_LAUNCHER=ccache +endif +.PHONY: build-editable clean deps-build deps-run build-wheel deps-docs docs docs-serve +build-editable: + git diff --exit-code --quiet '*/src/**' || (echo "Uncommitted changes in */src"; exit 1) + pip install --no-build-isolation --check-build-dependencies -Cbuild-dir=build --no-deps -t . -U -v . $(BUILD_CMAKE_FLAGS) + git restore '*/src/**' +test: + pytest -k "not perf" -n=3 + pytest -k "perf" -n=0 --cov-append +clean: + git clean -Xdf +deps-build: + pip install toml + python -c 'import toml; c=toml.load("pyproject.toml"); print("\0".join(c["build-system"]["requires"] + ["cmake>=" + c["tool"]["scikit-build"]["cmake"]["minimum-version"]]), end="")' | xargs -0 pip install ninja +deps-run: + pip install toml + python -c 'import toml; c=toml.load("pyproject.toml"); print("\0".join(c["project"]["dependencies"] + c["project"]["optional-dependencies"]["dev"]), end="")' | xargs -0 pip install +build-wheel: + pip install build + python -m build -n -w $(BUILD_CMAKE_FLAGS) +deps-docs: + cd docs && pip install -r requirements.txt +docs: + cd docs && PYTHONPATH=. pydoc-markdown --build --site-dir=../../../dist/site +docs-serve: docs + python -m http.server -d dist/site diff --git a/README.rst b/README.rst index fc6059c..96f4b4a 100644 --- a/README.rst +++ b/README.rst @@ -59,8 +59,8 @@ See also `NumCu `_, a minimal stand-alone Pytho External Projects ~~~~~~~~~~~~~~~~~ -For integration into Python, C++, CUDA, CMake, and general SWIG projects, see `the external project documentation `_. -Full and explicit example modules using the `CPython API `_ and `SWIG `_ are also provided. +For integration into Python, C++, CUDA, CMake, pybind11, and general SWIG projects, see `the external project documentation `_. +Full and explicit example modules using the `CPython API `_, `pybind11 API `_, and `SWIG `_ are also provided. Contributing ~~~~~~~~~~~~ @@ -83,7 +83,7 @@ Copyright: :target: https://github.com/AMYPAD/CuVec/blob/main/LICENCE .. |Tests| image:: https://img.shields.io/github/actions/workflow/status/AMYPAD/CuVec/test.yml?branch=main&logo=GitHub :target: https://github.com/AMYPAD/CuVec/actions -.. |Downloads| image:: https://img.shields.io/pypi/dm/cuvec.svg?logo=pypi&logoColor=white&label=PyPI%20downloads +.. |Downloads| image:: https://img.shields.io/pypi/dm/cuvec?logo=pypi&logoColor=white :target: https://pypi.org/project/cuvec .. |Coverage| image:: https://codecov.io/gh/AMYPAD/CuVec/branch/main/graph/badge.svg :target: https://codecov.io/gh/AMYPAD/CuVec diff --git a/cuvec/CMakeLists.txt b/cuvec/CMakeLists.txt index 727287d..6540096 100644 --- a/cuvec/CMakeLists.txt +++ b/cuvec/CMakeLists.txt @@ -1,6 +1,6 @@ cmake_minimum_required(VERSION 3.24 FATAL_ERROR) if(NOT DEFINED SKBUILD_PROJECT_VERSION) - set(SKBUILD_PROJECT_VERSION 2 CACHE STRING "version" FORCE) + set(SKBUILD_PROJECT_VERSION 4 CACHE STRING "version" FORCE) endif() string(REGEX REPLACE [[([0-9]+)\.([0-9]+)\.([0-9]+).*]] [[\1.\2.\3]] SKBUILD_PROJECT_VERSION "${SKBUILD_PROJECT_VERSION}") project(cuvec LANGUAGES C CXX VERSION "${SKBUILD_PROJECT_VERSION}") @@ -13,7 +13,7 @@ if(NOT DEFINED CMAKE_CUDA_ARCHITECTURES) set(CMAKE_CUDA_ARCHITECTURES native CACHE STRING "CUDA arch" FORCE) endif() cmake_policy(SET CMP0104 NEW) # CMAKE_CUDA_ARCHITECTURES -find_package(Python COMPONENTS Interpreter Development.Module REQUIRED) # NumPy +find_package(Python COMPONENTS Interpreter Development.Module REQUIRED) if(NOT CUVEC_CUDA_OPTIONAL) find_package(CUDAToolkit REQUIRED) enable_language(CUDA) @@ -31,6 +31,7 @@ else() endif() endif() if(SKBUILD) + find_package(pybind11 CONFIG) find_package(SWIG 4.0) if(SWIG_FOUND) include(${SWIG_USE_FILE}) @@ -49,45 +50,77 @@ message(STATUS "Build type: ${CMAKE_BUILD_TYPE}") option(CUVEC_DEBUG "Print out CUDA malloc & free operations" OFF) if(CUVEC_DEBUG) add_compile_definitions(CUVEC_DEBUG) + message(STATUS "CuVec debugging: TRUE") +else() + message(STATUS "CuVec debugging: FALSE") endif(CUVEC_DEBUG) -message(STATUS "CuVec debugging: ${CUVEC_DEBUG}") set(${CMAKE_PROJECT_NAME}_INCLUDE_DIRS "${CMAKE_CURRENT_LIST_DIR}/include/") # / suffix important install(DIRECTORY "${${CMAKE_PROJECT_NAME}_INCLUDE_DIRS}" DESTINATION ${CMAKE_PROJECT_NAME}/include) -# main project +# cpython extension -file(GLOB SRC LIST_DIRECTORIES false "src/*.cu") +file(GLOB SRC LIST_DIRECTORIES false "src/cpython*.cu") include_directories(src) -include_directories(${Python3_INCLUDE_DIRS}) +include_directories(${Python_INCLUDE_DIRS}) if(SKBUILD) - python_add_library(${PROJECT_NAME} MODULE WITH_SOABI ${SRC}) + python_add_library(${PROJECT_NAME}_cpython MODULE WITH_SOABI ${SRC}) else() - add_library(${PROJECT_NAME} SHARED ${SRC}) + add_library(${PROJECT_NAME}_cpython SHARED ${SRC}) endif() -add_library(AMYPAD::${PROJECT_NAME} ALIAS ${PROJECT_NAME}) -target_include_directories(${PROJECT_NAME} PUBLIC +add_library(AMYPAD::${PROJECT_NAME}_cpython ALIAS ${PROJECT_NAME}_cpython) +target_include_directories(${PROJECT_NAME}_cpython PUBLIC "$" "$") if(CUDAToolkit_FOUND) - target_link_libraries(${PROJECT_NAME} PRIVATE CUDA::cudart_static) + target_link_libraries(${PROJECT_NAME}_cpython PRIVATE CUDA::cudart_static) else() set_source_files_properties(${SRC} PROPERTIES LANGUAGE CXX) + target_link_libraries(${PROJECT_NAME}_cpython PRIVATE) endif() -set_target_properties(${PROJECT_NAME} PROPERTIES +set_target_properties(${PROJECT_NAME}_cpython PROPERTIES CXX_STANDARD 11 VERSION ${CMAKE_PROJECT_VERSION} SOVERSION ${CMAKE_PROJECT_VERSION_MAJOR} - INTERFACE_${PROJECT_NAME}_MAJOR_VERSION ${CMAKE_PROJECT_VERSION_MAJOR}) -set_property(TARGET ${PROJECT_NAME} APPEND PROPERTY COMPATIBLE_INTERFACE_STRING ${PROJECT_NAME}_MAJOR_VERSION) -install(TARGETS ${PROJECT_NAME} EXPORT ${PROJECT_NAME}Targets + INTERFACE_${PROJECT_NAME}_cpython_MAJOR_VERSION ${CMAKE_PROJECT_VERSION_MAJOR}) +set_property(TARGET ${PROJECT_NAME}_cpython APPEND PROPERTY COMPATIBLE_INTERFACE_STRING ${PROJECT_NAME}_cpython_MAJOR_VERSION) +install(TARGETS ${PROJECT_NAME}_cpython EXPORT ${PROJECT_NAME}Targets INCLUDES DESTINATION ${CMAKE_PROJECT_NAME}/include LIBRARY DESTINATION ${CMAKE_PROJECT_NAME}) install(EXPORT ${PROJECT_NAME}Targets FILE AMYPAD${PROJECT_NAME}Targets.cmake NAMESPACE AMYPAD:: DESTINATION ${CMAKE_PROJECT_NAME}/cmake) -# alternative swvec module +# alternative pybind11 extension + +if(pybind11_FOUND AND SKBUILD) + file(GLOB SRC LIST_DIRECTORIES false "src/pybind11*.cu") + # include_directories(src) + # include_directories(${Python_INCLUDE_DIRS}) + python_add_library(${PROJECT_NAME}_pybind11 MODULE WITH_SOABI ${SRC}) + add_library(AMYPAD::${PROJECT_NAME}_pybind11 ALIAS ${PROJECT_NAME}_pybind11) + target_include_directories(${PROJECT_NAME}_pybind11 PUBLIC + "$" + "$") + if(CUDAToolkit_FOUND) + target_link_libraries(${PROJECT_NAME}_pybind11 PRIVATE pybind11::headers CUDA::cudart_static) + else() + set_source_files_properties(${SRC} PROPERTIES LANGUAGE CXX) + target_link_libraries(${PROJECT_NAME}_pybind11 PRIVATE pybind11::headers) + endif() + + set_target_properties(${PROJECT_NAME}_pybind11 PROPERTIES + CXX_STANDARD 11 + VERSION ${CMAKE_PROJECT_VERSION} SOVERSION ${CMAKE_PROJECT_VERSION_MAJOR} + INTERFACE_pybind11_MAJOR_VERSION ${CMAKE_PROJECT_VERSION_MAJOR}) + set_property(TARGET ${PROJECT_NAME}_pybind11 APPEND PROPERTY COMPATIBLE_INTERFACE_STRING pybind11_MAJOR_VERSION) + install(TARGETS ${PROJECT_NAME}_pybind11 + INCLUDES DESTINATION ${CMAKE_PROJECT_NAME}/include + LIBRARY DESTINATION ${CMAKE_PROJECT_NAME}) +endif() + +# alternative swig extension + if(SWIG_FOUND AND SKBUILD) if(CUDAToolkit_FOUND) include_directories(${CUDAToolkit_INCLUDE_DIRS}) @@ -121,7 +154,12 @@ if(SWIG_FOUND AND SKBUILD) LIBRARY DESTINATION ${CMAKE_PROJECT_NAME}) endif() -add_subdirectory(src/example_mod) +# example projects + +add_subdirectory(src/example_cpython) +if(pybind11_FOUND AND SKBUILD) + add_subdirectory(src/example_pybind11) +endif() add_subdirectory(src/example_swig) # install project diff --git a/cuvec/__init__.py b/cuvec/__init__.py index c738722..e211d1d 100644 --- a/cuvec/__init__.py +++ b/cuvec/__init__.py @@ -15,17 +15,8 @@ __version__ = get_version(root="../..", relative_to=__file__) except (ImportError, LookupError): __version__ = "UNKNOWN" -__all__ = [ - # config - 'cmake_prefix', 'include_path', - # classes - 'CuVec', - # functions - 'dev_set', 'dev_sync', 'cu_copy', 'cu_zeros', - 'copy', 'asarray', - 'zeros', 'ones', 'zeros_like', 'ones_like', - # data - 'typecodes', 'vec_types'] # yapf: disable +# config +__all__ = ['cmake_prefix', 'include_path'] try: # py<3.9 import importlib_resources as resources @@ -33,24 +24,20 @@ from importlib import resources # type: ignore # yapf: disable try: - from .cuvec import dev_set, dev_sync + from .cuvec_cpython import dev_set, dev_sync except ImportError as err: # pragma: no cover from warnings import warn warn(str(err), UserWarning) -else: - from .pycuvec import ( - CuVec, - asarray, - copy, - cu_copy, - cu_zeros, - ones, - ones_like, - typecodes, - vec_types, - zeros, - zeros_like, - ) +else: # backwards compatibility: import from .cpython + from .cpython import CuVec, asarray, copy, ones, ones_like, typecodes, zeros, zeros_like + __all__ += [ + # classes + 'CuVec', + # functions + 'dev_set', 'dev_sync', 'copy', 'asarray', + 'zeros', 'ones', 'zeros_like', 'ones_like', + # data + 'typecodes'] # yapf: disable p = resources.files('cuvec').resolve() # for C++/CUDA/SWIG includes diff --git a/cuvec/_common.py b/cuvec/_common.py deleted file mode 100644 index 4112eff..0000000 --- a/cuvec/_common.py +++ /dev/null @@ -1,33 +0,0 @@ -"""Common helpers for pycuvec & swigcuvec modules.""" -import array -from typing import Sequence as Seq -from typing import Union - -Shape = Union[Seq[int], int] -# u: non-standard np.dype('S2'); l/L: inconsistent between `array` and `numpy` -typecodes = ''.join(i for i in array.typecodes if i not in "ulL") - - -def _generate_helpers(zeros, CuVec): - def ones(shape: Shape, dtype="float32") -> CuVec: - """ - Returns a `CuVec` view of a new `numpy.ndarray` - of the specified shape and data type (equivalent of `numpy.ones`). - """ - res = zeros(shape, dtype) - res[:] = 1 - return res - - def zeros_like(arr) -> CuVec: - """ - Returns `zeros(arr.shape, arr.dtype)`. - """ - return zeros(arr.shape, arr.dtype) - - def ones_like(arr) -> CuVec: - """ - Returns `ones(arr.shape, arr.dtype)`. - """ - return ones(arr.shape, arr.dtype) - - return ones, zeros_like, ones_like diff --git a/cuvec/_utils.py b/cuvec/_utils.py new file mode 100644 index 0000000..0cf6a6f --- /dev/null +++ b/cuvec/_utils.py @@ -0,0 +1,91 @@ +"""Common helpers for cuvec.{cpython,pybind11,swig} modules.""" +import array +import re +from abc import ABC, abstractmethod +from typing import Any, Dict +from typing import Sequence as Seq +from typing import Union + +import numpy as np + +Shape = Union[Seq[int], int] +# u: non-standard np.dype('S2'); l/L: inconsistent between `array` and `numpy` +typecodes = ''.join(i for i in array.typecodes if i not in "ulL") + + +class CVector(ABC): + """Thin wrapper around `CuVec`. Always takes ownership.""" + vec_types: Dict[np.dtype, Any] + RE_CUVEC_TYPE: re.Pattern + + def __init__(self, typechar: str): + self.typechar = typechar + + @property + @abstractmethod + def shape(self) -> tuple: + pass # pragma: no cover + + @property + @abstractmethod + def address(self) -> int: + pass # pragma: no cover + + @property + def __array_interface__(self) -> Dict[str, Any]: + return { + 'shape': self.shape, 'typestr': np.dtype(self.typechar).str, + 'data': (self.address, False), 'version': 3} + + __cuda_array_interface__ = __array_interface__ + + def __repr__(self) -> str: + return f"{type(self).__name__}('{self.typechar}', {self.shape})" + + def __str__(self) -> str: + return f"{np.dtype(self.typechar)}{self.shape} at 0x{self.address:x}" + + @classmethod + def zeros(cls, shape: Shape, dtype="float32"): + """Returns a new Vector of the specified shape and data type.""" + return cls.vec_types[np.dtype(dtype)](shape) + + @classmethod + def copy(cls, arr): + """Returns a new Vector with data copied from the specified `arr`.""" + res = cls.zeros(arr.shape, arr.dtype) + np.asarray(res).flat = arr.flat + return res + + @classmethod + def is_instance(cls, arr): + return isinstance(arr, cls) or type(arr).__name__ == cls.__name__ + + @classmethod + def is_raw_cuvec(cls, arr): + return cls.RE_CUVEC_TYPE.match(str(arr)) + + +def _generate_helpers(zeros, CuVec): + def ones(shape: Shape, dtype="float32") -> CuVec: + """ + Returns a `CuVec` view of a new `numpy.ndarray` + of the specified shape and data type (equivalent of `numpy.ones`). + """ + res = zeros(shape, dtype) + res[:] = 1 + return res + + def zeros_like(arr) -> CuVec: + """ + Returns `zeros(arr.shape, arr.dtype)`. + """ + return zeros(arr.shape, arr.dtype) + + def ones_like(arr) -> CuVec: + """ + Returns `ones(arr.shape, arr.dtype)`. + """ + return ones(arr.shape, arr.dtype) + + return ones, zeros_like, ones_like diff --git a/cuvec/pycuvec.py b/cuvec/cpython.py similarity index 92% rename from cuvec/pycuvec.py rename to cuvec/cpython.py index f1644e8..5ae4625 100644 --- a/cuvec/pycuvec.py +++ b/cuvec/cpython.py @@ -1,4 +1,4 @@ -"""Thin wrappers around `cuvec` C++/CUDA module""" +"""Thin wrappers around `cuvec_cpython` C++/CUDA module""" import logging from collections.abc import Sequence from textwrap import dedent @@ -6,8 +6,11 @@ import numpy as np -from . import cuvec as cu -from ._common import Shape, _generate_helpers, typecodes +from . import cuvec_cpython as cu +from ._utils import Shape, _generate_helpers, typecodes + +__all__ = [ + 'CuVec', 'zeros', 'ones', 'zeros_like', 'ones_like', 'copy', 'asarray', 'Shape', 'typecodes'] log = logging.getLogger(__name__) vec_types = { @@ -54,7 +57,7 @@ def is_raw_cuvec(cuvec): This is needed since conversely `isinstance(cuvec, CuVec)` may be `False` due to external libraries - `#include "pycuvec.cuh"` making a distinct type object. + `#include "cuvec_cpython.cuh"` making a distinct type object. """ return isinstance(cuvec, _PyCuVec_types) or str(type(cuvec)) in _PyCuVec_types_s diff --git a/cuvec/include/cuvec.cuh b/cuvec/include/cuvec.cuh index 5317690..b1847e0 100644 --- a/cuvec/include/cuvec.cuh +++ b/cuvec/include/cuvec.cuh @@ -16,11 +16,12 @@ #ifndef CUVEC_DISABLE_CUDA #include "cuda_runtime.h" #endif -#include // fprintf -#include // std::size_t, std::malloc, std::free -#include // std::numeric_limits -#include // std::bad_alloc -#include // std::vector +#include // fprintf +#include // std::size_t, std::malloc, std::free +#include // std::numeric_limits +#include // std::bad_alloc +#include // std::length_error +#include // std::vector #ifndef CUVEC_DISABLE_CUDA namespace cuvec { @@ -88,23 +89,50 @@ template bool operator!=(const CuAlloc &, const CuAlloc template using CuVec = std::vector>; -template struct SwigCuVec { +/// extension helpers +#ifndef _CUVEC_HALF +#ifndef CUVEC_DISABLE_CUDA +#include "cuda_fp16.h" // __half +#define _CUVEC_HALF __half +#else // CUVEC_DISABLE_CUDA +#ifdef __fp16 +#define _CUVEC_HALF __fp16 +#endif // __fp16 +#endif // CUVEC_DISABLE_CUDA +#endif // _CUVEC_HALF + +/// pybind11 helpers +template struct NDCuVec { CuVec vec; std::vector shape; + NDCuVec() = default; + NDCuVec(const std::vector &shape) : shape(shape) { + size_t size = 1; + for (auto &i : shape) size *= i; + vec.resize(size); + } + void reshape(const std::vector &shape) { + size_t size = 1; + for (auto &i : shape) size *= i; + if (size != vec.size()) throw std::length_error("reshape: size mismatch"); + this->shape = shape; + } + ~NDCuVec() { + vec.clear(); + vec.shrink_to_fit(); + shape.clear(); + shape.shrink_to_fit(); + } }; + +/// SWIG helpers +template using SwigCuVec = NDCuVec; template SwigCuVec *SwigCuVec_new(std::vector shape) { - SwigCuVec *self = new SwigCuVec; - self->shape = shape; - size_t size = 1; - for (auto &i : shape) size *= i; - self->vec.resize(size); + SwigCuVec *self = new SwigCuVec(shape); return self; } template void SwigCuVec_del(SwigCuVec *self) { - self->vec.clear(); - self->vec.shrink_to_fit(); - self->shape.clear(); - self->shape.shrink_to_fit(); + self->~NDCuVec(); delete self; } template T *SwigCuVec_data(SwigCuVec *self) { return self->vec.data(); } @@ -112,5 +140,8 @@ template size_t SwigCuVec_address(SwigCuVec *self) { return (size_t)SwigCuVec_data(self); } template std::vector SwigCuVec_shape(SwigCuVec *self) { return self->shape; } +template void SwigCuVec_reshape(SwigCuVec *self, const std::vector &shape) { + self->reshape(shape); +} #endif // _CUVEC_H_ diff --git a/cuvec/include/pycuvec.cuh b/cuvec/include/cuvec_cpython.cuh similarity index 94% rename from cuvec/include/pycuvec.cuh rename to cuvec/include/cuvec_cpython.cuh index d7843aa..8f86b60 100644 --- a/cuvec/include/pycuvec.cuh +++ b/cuvec/include/cuvec_cpython.cuh @@ -6,25 +6,15 @@ * PyCuVec *PyCuVec_deepcopy(PyCuVec *other); * PyTypeObject PyCuVec_tp.tp_obj; */ -#ifndef _PYCUVEC_H_ -#define _PYCUVEC_H_ +#ifndef _CUVEC_CPYTHON_H_ +#define _CUVEC_CPYTHON_H_ -#include "Python.h" -#ifndef _CUVEC_HALF -#ifndef CUVEC_DISABLE_CUDA -#include "cuda_fp16.h" // __half -#define _CUVEC_HALF __half -#else // CUVEC_DISABLE_CUDA -#ifdef __fp16 -#define _CUVEC_HALF __fp16 -#endif // __fp16 -#endif // CUVEC_DISABLE_CUDA -#endif // _CUVEC_HALF -#include "cuvec.cuh" // CuVec -#include // malloc, free -#include // std::stringstream -#include // typeid -#include // std::vector +#include "cuvec.cuh" // CuVec, _CUVEC_HALF +#include +#include // malloc, free +#include // std::stringstream +#include // typeid +#include // std::vector namespace cuvec { template struct PyType { @@ -235,7 +225,7 @@ template PyCuVec *PyCuVec_zeros(std::vector shape) { self->vec.resize(self->shape[0] * (self->strides[0] / sizeof(T))); return CUDA_PyErr() ? NULL : self; } -template PyCuVec *PyCuVec_zeros_like(PyCuVec *other) { +template PyCuVec *PyCuVec_zeros_like(const PyCuVec *other) { PyCuVec *self = PyCuVec_new(); if (!self) return NULL; self->vec.resize(other->vec.size()); @@ -244,7 +234,7 @@ template PyCuVec *PyCuVec_zeros_like(PyCuVec *other) { self->strides = other->strides; return self; } -template PyCuVec *PyCuVec_deepcopy(PyCuVec *other) { +template PyCuVec *PyCuVec_deepcopy(const PyCuVec *other) { PyCuVec *self = PyCuVec_new(); if (!self) return NULL; self->vec = other->vec; @@ -330,4 +320,4 @@ ASCUVEC(_CUVEC_HALF, e) ASCUVEC(float, f) ASCUVEC(double, d) -#endif // _PYCUVEC_H_ +#endif // _CUVEC_CPYTHON_H_ diff --git a/cuvec/include/cuvec_pybind11.cuh b/cuvec/include/cuvec_pybind11.cuh new file mode 100644 index 0000000..08e5886 --- /dev/null +++ b/cuvec/include/cuvec_pybind11.cuh @@ -0,0 +1,36 @@ +/** + * Python template header wrapping `NDCuVec`. Provides: + * PYBIND11_BIND_NDCUVEC(T, typechar); + */ +#ifndef _CUVEC_PYBIND11_H_ +#define _CUVEC_PYBIND11_H_ + +#include "cuvec.cuh" // NDCuVec +#include // pybind11, PYBIND11_MAKE_OPAQUE +#include // std::vector + +PYBIND11_MAKE_OPAQUE(std::vector); +PYBIND11_MAKE_OPAQUE(NDCuVec); +PYBIND11_MAKE_OPAQUE(NDCuVec); +PYBIND11_MAKE_OPAQUE(NDCuVec); +PYBIND11_MAKE_OPAQUE(NDCuVec); +PYBIND11_MAKE_OPAQUE(NDCuVec); +PYBIND11_MAKE_OPAQUE(NDCuVec); +PYBIND11_MAKE_OPAQUE(NDCuVec); +PYBIND11_MAKE_OPAQUE(NDCuVec); +PYBIND11_MAKE_OPAQUE(NDCuVec); +#ifdef _CUVEC_HALF +PYBIND11_MAKE_OPAQUE(NDCuVec<_CUVEC_HALF>); +#endif +PYBIND11_MAKE_OPAQUE(NDCuVec); +PYBIND11_MAKE_OPAQUE(NDCuVec); + +#define PYBIND11_BIND_NDCUVEC(T, typechar) \ + pybind11::class_>(m, PYBIND11_TOSTRING(NDCuVec_##typechar)) \ + .def(pybind11::init<>()) \ + .def(pybind11::init>()) \ + .def("reshape", &NDCuVec::reshape) \ + .def("shape", [](const NDCuVec &v) { return v.shape; }) \ + .def("address", [](NDCuVec &v) { return (size_t)v.vec.data(); }) + +#endif // _CUVEC_PYBIND11_H_ diff --git a/cuvec/pybind11.py b/cuvec/pybind11.py new file mode 100644 index 0000000..578e206 --- /dev/null +++ b/cuvec/pybind11.py @@ -0,0 +1,163 @@ +""" +Thin wrappers around `cuvec_pybind11` C++/CUDA module + +A pybind11-driven equivalent of the CPython Extension API-driven `cpython.py` +""" +import logging +import re +from collections.abc import Sequence +from functools import partial +from textwrap import dedent +from typing import Any, Dict, Optional + +import numpy as np + +from . import cuvec_pybind11 as cu # type: ignore [attr-defined] # yapf: disable +from ._utils import CVector, Shape, _generate_helpers, typecodes + +__all__ = [ + 'CuVec', 'zeros', 'ones', 'zeros_like', 'ones_like', 'copy', 'asarray', 'retarray', 'Shape', + 'typecodes'] + +log = logging.getLogger(__name__) +if hasattr(cu, 'NDCuVec_e'): + typecodes += 'e' + + +class Pybind11Vector(CVector): + RE_CUVEC_TYPE = re.compile(r"<.*NDCuVec_(.) object at 0x\w+>") + + def __init__(self, typechar: str, shape: Shape, cuvec=None): + """ + Args: + typechar(char) + shape(tuple(int)) + cuvec(NDCuVec): if given, `typechar` and `shape` are ignored + """ + if cuvec is None: + shape = cu.Shape(shape if isinstance(shape, Sequence) else (shape,)) + cuvec = getattr(cu, f'NDCuVec_{typechar}')(shape) + else: + typechar = self.is_raw_cuvec(cuvec).group(1) + self.cuvec = cuvec + super().__init__(typechar) + + @property + def shape(self) -> tuple: + return tuple(self.cuvec.shape()) + + @shape.setter + def shape(self, shape: Shape): + shape = cu.Shape(shape if isinstance(shape, Sequence) else (shape,)) + self.cuvec.reshape(shape) + + @property + def address(self) -> int: + return self.cuvec.address() + + +Pybind11Vector.vec_types = {np.dtype(c): partial(Pybind11Vector, c) for c in typecodes} + + +class CuVec(np.ndarray): + """ + A `numpy.ndarray` compatible view with a `cuvec` member containing the + underlying `Pybind11Vector` object (for use in pybind11 API function calls). + """ + def __new__(cls, arr): + """arr: `cuvec.pybind11.CuVec`, raw `Pybind11Vector`, or `numpy.ndarray`""" + if Pybind11Vector.is_instance(arr): + log.debug("wrap pyraw %s", type(arr)) + obj = np.asarray(arr).view(cls) + obj._vec = arr + obj.cuvec = arr.cuvec + return obj + if isinstance(arr, CuVec) and hasattr(arr, '_vec'): + log.debug("new view") + obj = np.asarray(arr).view(cls) + obj._vec = arr._vec + obj.cuvec = arr._vec.cuvec + return obj + if isinstance(arr, np.ndarray): + log.debug("copy") + return copy(arr) + raise NotImplementedError( + dedent("""\ + Not intended for explicit construction + (do not do `cuvec.pybind11.CuVec((42, 1337))`; + instead use `cuvec.pybind11.zeros((42, 137))`""")) + + @property + def __cuda_array_interface__(self) -> Dict[str, Any]: + if not hasattr(self, 'cuvec'): + raise AttributeError( + dedent("""\ + `numpy.ndarray` object has no attribute `cuvec`: + try using `cuvec.asarray()` first.""")) + return self._vec.__cuda_array_interface__ + + def resize(self, new_shape: Shape): + """Change shape (but not size) of array in-place.""" + self._vec.shape = new_shape + super().resize(new_shape, refcheck=False) + + +def zeros(shape: Shape, dtype="float32") -> CuVec: + """ + Returns a `cuvec.pybind11.CuVec` view of a new `numpy.ndarray` + of the specified shape and data type (`cuvec` equivalent of `numpy.zeros`). + """ + return CuVec(Pybind11Vector.zeros(shape, dtype)) + + +ones, zeros_like, ones_like = _generate_helpers(zeros, CuVec) + + +def copy(arr) -> CuVec: + """ + Returns a `cuvec.pybind11.CuVec` view of a new `numpy.ndarray` + with data copied from the specified `arr` + (`cuvec` equivalent of `numpy.copy`). + """ + return CuVec(Pybind11Vector.copy(arr)) + + +def asarray(arr, dtype=None, order=None, ownership: str = 'warning') -> CuVec: + """ + Returns a `cuvec.pybind11.CuVec` view of `arr`, avoiding memory copies if possible. + (`cuvec` equivalent of `numpy.asarray`). + + Args: + ownership: logging level if `is_raw_cuvec(arr)`. + WARNING: `asarray()` should not be used on an existing reference, e.g.: + >>> res = asarray(some_pybind11_api_func(..., output=getattr(out, 'cuvec', None))) + `res.cuvec` and `out.cuvec` are now the same + yet garbage collected separately (dangling ptr). + Instead, use the `retarray` helper: + >>> raw = some_pybind11_api_func(..., output=getattr(out, 'cuvec', None)) + >>> res = retarray(raw, out) + NB: `asarray()`/`retarray()` are safe if the raw cuvec was created in C++, e.g.: + >>> res = retarray(some_pybind11_api_func(..., output=None)) + """ + if Pybind11Vector.is_raw_cuvec(arr): + ownership = ownership.lower() + if ownership in {'critical', 'fatal', 'error'}: + raise IOError("Can't take ownership of existing cuvec (would create dangling ptr)") + getattr(log, ownership)("taking ownership") + arr = Pybind11Vector('', (), arr) + if not isinstance(arr, np.ndarray) and Pybind11Vector.is_instance(arr): + res = CuVec(arr) + if dtype is None or res.dtype == np.dtype(dtype): + return CuVec(np.asanyarray(res, order=order)) + return CuVec(np.asanyarray(arr, dtype=dtype, order=order)) + + +def retarray(raw, out: Optional[CuVec] = None): + """ + Returns `out if hasattr(out, 'cuvec') else asarray(raw, ownership='debug')`. + See `asarray` for explanation. + Args: + raw: a raw CuVec (returned by C++/pybind11 function). + out: preallocated output array. + """ + return out if hasattr(out, 'cuvec') else asarray(raw, ownership='debug') diff --git a/cuvec/src/cuvec.cu b/cuvec/src/cpython.cu similarity index 95% rename from cuvec/src/cuvec.cu rename to cuvec/src/cpython.cu index 99e2f36..3d77883 100644 --- a/cuvec/src/cuvec.cu +++ b/cuvec/src/cpython.cu @@ -5,8 +5,9 @@ * * Copyright (2021) Casper da Costa-Luis */ -#include "Python.h" -#include "pycuvec.cuh" // PyCuVec, PyCuVec_tp +#include "cuvec_cpython.cuh" // PyCuVec, PyCuVec_tp +#include + /** functions */ /// required before accessing on host static PyObject *dev_sync(PyObject *self, PyObject *args) { @@ -37,11 +38,11 @@ static PyMethodDef cuvec_methods[] = { /** module */ static struct PyModuleDef cuvec_module = { PyModuleDef_HEAD_INIT, - "cuvec", // module + "cuvec_cpython", // module "CUDA unified memory with python array buffer and C++ std::vector interfaces.", -1, // module keeps state in global variables cuvec_methods}; -PyMODINIT_FUNC PyInit_cuvec(void) { +PyMODINIT_FUNC PyInit_cuvec_cpython(void) { Py_Initialize(); // import_array(); // load NumPy functionality diff --git a/cuvec/src/example_mod/CMakeLists.txt b/cuvec/src/example_cpython/CMakeLists.txt similarity index 93% rename from cuvec/src/example_mod/CMakeLists.txt rename to cuvec/src/example_cpython/CMakeLists.txt index 9366310..1d418dd 100644 --- a/cuvec/src/example_mod/CMakeLists.txt +++ b/cuvec/src/example_cpython/CMakeLists.txt @@ -1,7 +1,7 @@ -project(example_mod) +project(example_cpython) file(GLOB SRC LIST_DIRECTORIES false "*.cu") -include_directories(${Python3_INCLUDE_DIRS}) +include_directories(${Python_INCLUDE_DIRS}) if(SKBUILD) python_add_library(${PROJECT_NAME} MODULE WITH_SOABI ${SRC}) diff --git a/cuvec/src/example_mod/example_mod.cu b/cuvec/src/example_cpython/example_mod.cu similarity index 85% rename from cuvec/src/example_mod/example_mod.cu rename to cuvec/src/example_cpython/example_mod.cu index fc59f05..df3bec4 100644 --- a/cuvec/src/example_mod/example_mod.cu +++ b/cuvec/src/example_cpython/example_mod.cu @@ -4,7 +4,7 @@ * Copyright (2021) Casper da Costa-Luis */ #include "Python.h" -#include "pycuvec.cuh" // PyCuVec +#include "cuvec_cpython.cuh" // PyCuVec #ifdef CUVEC_DISABLE_CUDA #include // std::chrono #else @@ -89,12 +89,12 @@ static PyMethodDef example_methods[] = { }; /** module */ -static struct PyModuleDef example_mod = {PyModuleDef_HEAD_INIT, - "example_mod", // module - "Example external module.", - -1, // module keeps state in global variables - example_methods}; -PyMODINIT_FUNC PyInit_example_mod(void) { +static struct PyModuleDef example_cpython = {PyModuleDef_HEAD_INIT, + "example_cpython", // module + "Example external module.", + -1, // module keeps state in global variables + example_methods}; +PyMODINIT_FUNC PyInit_example_cpython(void) { Py_Initialize(); - return PyModule_Create(&example_mod); + return PyModule_Create(&example_cpython); } diff --git a/cuvec/src/example_pybind11/CMakeLists.txt b/cuvec/src/example_pybind11/CMakeLists.txt new file mode 100644 index 0000000..998bbed --- /dev/null +++ b/cuvec/src/example_pybind11/CMakeLists.txt @@ -0,0 +1,25 @@ +project(example_pybind11) +file(GLOB SRC LIST_DIRECTORIES false "*.cu") + +if(CUDAToolkit_FOUND) + include_directories(${CUDAToolkit_INCLUDE_DIRS}) +endif() +python_add_library(${PROJECT_NAME} MODULE WITH_SOABI ${SRC}) +target_include_directories(${PROJECT_NAME} PUBLIC + "$" + "$") +if(CUDAToolkit_FOUND) + target_link_libraries(${PROJECT_NAME} PRIVATE pybind11::headers CUDA::cudart_static) +else() + set_source_files_properties(${SRC} PROPERTIES LANGUAGE CXX) + target_link_libraries(${PROJECT_NAME} PRIVATE pybind11::headers) +endif() + +set_target_properties(${PROJECT_NAME} PROPERTIES + CXX_STANDARD 11 + VERSION ${CMAKE_PROJECT_VERSION} SOVERSION ${CMAKE_PROJECT_VERSION_MAJOR} + INTERFACE_${PROJECT_NAME}_MAJOR_VERSION ${CMAKE_PROJECT_VERSION_MAJOR}) +set_property(TARGET ${PROJECT_NAME} APPEND PROPERTY COMPATIBLE_INTERFACE_STRING ${PROJECT_NAME}_MAJOR_VERSION) +install(TARGETS ${PROJECT_NAME} + INCLUDES DESTINATION ${CMAKE_PROJECT_NAME}/include + LIBRARY DESTINATION ${CMAKE_PROJECT_NAME}) diff --git a/cuvec/src/example_pybind11/example_pybind11.cu b/cuvec/src/example_pybind11/example_pybind11.cu new file mode 100644 index 0000000..4137ec3 --- /dev/null +++ b/cuvec/src/example_pybind11/example_pybind11.cu @@ -0,0 +1,74 @@ +/** + * Example external pybind11 extension module using CuVec. + * + * Copyright (2021) Casper da Costa-Luis + */ +#include "cuvec_pybind11.cuh" // NDCuVec +#include // pybind11, PYBIND11_MODULE +#include // std::length_error +#ifdef CUVEC_DISABLE_CUDA +#include // std::chrono +#else +/// dst = src + 1 +__global__ void _d_incr(float *dst, float *src, int X, int Y) { + int x = threadIdx.x + blockDim.x * blockIdx.x; + if (x >= X) return; + int y = threadIdx.y + blockDim.y * blockIdx.y; + if (y >= Y) return; + dst[y * X + x] = src[y * X + x] + 1; +} +#endif // CUVEC_DISABLE_CUDA +NDCuVec *increment2d_f(NDCuVec &src, NDCuVec *output, bool timing) { + auto &N = src.shape; + if (N.size() != 2) throw std::length_error("`src` must be 2D"); + +#ifndef CUVEC_DISABLE_CUDA + cudaEvent_t eStart, eAlloc, eKern; + cudaEventCreate(&eStart); + cudaEventCreate(&eAlloc); + cudaEventCreate(&eKern); + cudaEventRecord(eStart); +#else + auto eStart = std::chrono::steady_clock::now(); +#endif + + if (!output) + output = new NDCuVec(N); + else if (N != output->shape) + throw std::length_error("`output` must be same shape as `src`"); + +#ifndef CUVEC_DISABLE_CUDA + cudaEventRecord(eAlloc); + dim3 thrds((N[1] + 31) / 32, (N[0] + 31) / 32); + dim3 blcks(32, 32); + _d_incr<<>>(output->vec.data(), src.vec.data(), N[1], N[0]); + cuvec::HandleError(cudaGetLastError(), __FILE__, __LINE__); + // cudaDeviceSynchronize(); + cudaEventRecord(eKern); + cudaEventSynchronize(eKern); + float alloc_ms, kernel_ms; + cudaEventElapsedTime(&alloc_ms, eStart, eAlloc); + cudaEventElapsedTime(&kernel_ms, eAlloc, eKern); +// fprintf(stderr, "%.3f ms, %.3f ms\n", alloc_ms, kernel_ms); +#else + auto eAlloc = std::chrono::steady_clock::now(); + for (size_t i = 0; i < src.vec.size(); i++) output->vec[i] = src.vec[i] + 1; + auto eKern = std::chrono::steady_clock::now(); + double alloc_ms = std::chrono::duration(eAlloc - eStart).count(); + double kernel_ms = std::chrono::duration(eKern - eAlloc).count(); +// fprintf(stderr, "%.3lf ms, %.3lf ms\n", alloc_ms, kernel_ms); +#endif + if (timing) { + // hack: store times in first two elements of output + output->vec[0] = alloc_ms; + output->vec[1] = kernel_ms; + } + return output; +} + +using namespace pybind11::literals; // _a +PYBIND11_MODULE(example_pybind11, m) { + m.doc() = "Example external module."; + m.def("increment2d_f", &increment2d_f, "Returns: alloc_ms, kernel_ms, src + 1.", "src"_a, + "output"_a = nullptr, "timing"_a = false); +} diff --git a/cuvec/src/example_swig/CMakeLists.txt b/cuvec/src/example_swig/CMakeLists.txt index ca9e78c..5935a9b 100644 --- a/cuvec/src/example_swig/CMakeLists.txt +++ b/cuvec/src/example_swig/CMakeLists.txt @@ -2,8 +2,6 @@ project(example_swig) file(GLOB SRC LIST_DIRECTORIES false "*.cu") file(GLOB ISRC LIST_DIRECTORIES false "*.i") -#include_directories(${Python3_INCLUDE_DIRS}) - if(SWIG_FOUND) if(CUDAToolkit_FOUND) include_directories(${CUDAToolkit_INCLUDE_DIRS}) diff --git a/cuvec/src/pybind11.cu b/cuvec/src/pybind11.cu new file mode 100644 index 0000000..9fcd464 --- /dev/null +++ b/cuvec/src/pybind11.cu @@ -0,0 +1,29 @@ +/** + * Unifying Python/C++/CUDA memory. + * + * pybind11 opaque vector -> C++11 `std::vector` -> CUDA managed memory. + * + * Copyright (2024) Casper da Costa-Luis + */ +#include "cuvec_pybind11.cuh" // PYBIND11_BIND_NDCUVEC +#include // PYBIND11_MODULE +#include // pybind11::bind_vector + +PYBIND11_MODULE(cuvec_pybind11, m) { + m.doc() = "PyBind11 external module."; + pybind11::bind_vector>(m, "Shape"); + PYBIND11_BIND_NDCUVEC(signed char, b); + PYBIND11_BIND_NDCUVEC(unsigned char, B); + PYBIND11_BIND_NDCUVEC(char, c); + PYBIND11_BIND_NDCUVEC(short, h); + PYBIND11_BIND_NDCUVEC(unsigned short, H); + PYBIND11_BIND_NDCUVEC(int, i); + PYBIND11_BIND_NDCUVEC(unsigned int, I); + PYBIND11_BIND_NDCUVEC(long long, q); + PYBIND11_BIND_NDCUVEC(unsigned long long, Q); +#ifdef _CUVEC_HALF + PYBIND11_BIND_NDCUVEC(_CUVEC_HALF, e); +#endif + PYBIND11_BIND_NDCUVEC(float, f); + PYBIND11_BIND_NDCUVEC(double, d); +} diff --git a/cuvec/src/swvec.i b/cuvec/src/swvec.i index c022a32..526834b 100644 --- a/cuvec/src/swvec.i +++ b/cuvec/src/swvec.i @@ -9,19 +9,6 @@ } } -%{ -#ifndef _CUVEC_HALF -#ifndef CUVEC_DISABLE_CUDA -#include "cuda_fp16.h" // __half -#define _CUVEC_HALF __half -#else // CUVEC_DISABLE_CUDA -#ifdef __fp16 -#define _CUVEC_HALF __fp16 -#endif // __fp16 -#endif // CUVEC_DISABLE_CUDA -#endif // _CUVEC_HALF -%} - %include "cuvec.i" // SwigCuVec template SwigCuVec *SwigCuVec_new(std::vector shape); @@ -29,6 +16,7 @@ template void SwigCuVec_del(SwigCuVec *self); template T *SwigCuVec_data(SwigCuVec *self); template size_t SwigCuVec_address(SwigCuVec *self); template std::vector SwigCuVec_shape(SwigCuVec *self); +template void SwigCuVec_reshape(SwigCuVec *self, const std::vector &shape); %template(SwigCuVec_Shape) std::vector; %define MKCUVEC(T, typechar) @@ -38,6 +26,7 @@ template std::vector SwigCuVec_shape(SwigCuVec *self); %template(SwigCuVec_ ## typechar ## _data) SwigCuVec_data; %template(SwigCuVec_ ## typechar ## _address) SwigCuVec_address; %template(SwigCuVec_ ## typechar ## _shape) SwigCuVec_shape; +%template(SwigCuVec_ ## typechar ## _reshape) SwigCuVec_reshape; %enddef MKCUVEC(signed char, b) MKCUVEC(unsigned char, B) diff --git a/cuvec/swigcuvec.py b/cuvec/swig.py similarity index 54% rename from cuvec/swigcuvec.py rename to cuvec/swig.py index ba97aee..a1c2d02 100644 --- a/cuvec/swigcuvec.py +++ b/cuvec/swig.py @@ -1,7 +1,7 @@ """ Thin wrappers around `swvec` C++/CUDA module -A SWIG-driven equivalent of the CPython Extension API-driven `pycuvec.py` +A SWIG-driven equivalent of the CPython Extension API-driven `cpython.py` """ import logging import re @@ -12,48 +12,38 @@ import numpy as np -from . import swvec as sw # type: ignore # yapf: disable -from ._common import Shape, _generate_helpers, typecodes +from . import swvec as sw # type: ignore [attr-defined] # yapf: disable +from ._utils import CVector, Shape, _generate_helpers, typecodes + +__all__ = [ + 'CuVec', 'zeros', 'ones', 'zeros_like', 'ones_like', 'copy', 'asarray', 'retarray', 'Shape', + 'typecodes'] log = logging.getLogger(__name__) -RE_SWIG_TYPE = ("<.*SwigCuVec_(.); proxy of \s*\*' at 0x\w+>") -SWIG_TYPES = { - "signed char": 'b', - "unsigned char": 'B', - "char": 'c', - "short": 'h', - "unsigned short": 'H', - "int": 'i', - "unsigned int": 'I', - "long long": 'q', - "unsigned long long": 'Q', - "float": 'f', - "double": 'd'} # yapf: disable if hasattr(sw, 'SwigCuVec_e_new'): typecodes += 'e' - SWIG_TYPES["__half"] = 'e' -class SWIGVector: - def __init__(self, typechar: Optional[str], shape: Optional[Shape], cuvec=None): +class SWIGVector(CVector): + RE_CUVEC_TYPE = re.compile("<.*SwigCuVec_(.); proxy of \s*\*' at 0x\w+>") + + def __init__(self, typechar: str, shape: Shape, cuvec=None): """ - Thin wrapper around `SwigPyObject>`. Always takes ownership. Args: typechar(char) shape(tuple(int)) cuvec(SwigPyObject>): if given, `typechar` and `shape` are ignored """ - if cuvec is not None: - assert is_raw_cuvec(cuvec) - self.typechar = re.match(RE_SWIG_TYPE, str(cuvec)).group(1) # type: ignore - self.cuvec = cuvec - return - - self.typechar = typechar # type: ignore - self.cuvec = getattr( - sw, f'SwigCuVec_{typechar}_new')(shape if isinstance(shape, Sequence) else (shape,)) + if cuvec is None: + cuvec = getattr( + sw, + f'SwigCuVec_{typechar}_new')(shape if isinstance(shape, Sequence) else (shape,)) + else: + typechar = self.is_raw_cuvec(cuvec).group(1) + self.cuvec = cuvec + super().__init__(typechar) def __del__(self): getattr(sw, f'SwigCuVec_{self.typechar}_del')(self.cuvec) @@ -62,52 +52,17 @@ def __del__(self): def shape(self) -> tuple: return getattr(sw, f'SwigCuVec_{self.typechar}_shape')(self.cuvec) + @shape.setter + def shape(self, shape: Shape): + shape = shape if isinstance(shape, Sequence) else (shape,) + getattr(sw, f'SwigCuVec_{self.typechar}_reshape')(self.cuvec, shape) + @property def address(self) -> int: return getattr(sw, f'SwigCuVec_{self.typechar}_address')(self.cuvec) - @property - def __array_interface__(self) -> Dict[str, Any]: - return { - 'shape': self.shape, 'typestr': np.dtype(self.typechar).str, - 'data': (self.address, False), 'version': 3} - - @property - def __cuda_array_interface__(self) -> Dict[str, Any]: - return self.__array_interface__ - - def __repr__(self) -> str: - return f"{type(self).__name__}('{self.typechar}', {self.shape})" - - def __str__(self) -> str: - return f"{np.dtype(self.typechar)}{self.shape} at 0x{self.address:x}" - - -vec_types = {np.dtype(c): partial(SWIGVector, c) for c in typecodes} - -def cu_zeros(shape: Shape, dtype="float32"): - """ - Returns a new `SWIGVector` of the specified shape and data type. - """ - return vec_types[np.dtype(dtype)](shape) - - -def cu_copy(arr): - """ - Returns a new `SWIGVector` with data copied from the specified `arr`. - """ - res = cu_zeros(arr.shape, arr.dtype) - np.asarray(res).flat = arr.flat - return res - - -def is_raw_cuvec(arr): - return re.match(RE_SWIG_TYPE, str(arr)) - - -def is_raw_swvec(arr): - return isinstance(arr, SWIGVector) or type(arr).__name__ == "SWIGVector" +SWIGVector.vec_types = {np.dtype(c): partial(SWIGVector, c) for c in typecodes} class CuVec(np.ndarray): @@ -116,18 +71,18 @@ class CuVec(np.ndarray): underlying `SWIGVector` object (for use in CPython API function calls). """ def __new__(cls, arr): - """arr: `swigcuvec.CuVec`, raw `SWIGVector`, or `numpy.ndarray`""" - if is_raw_swvec(arr): + """arr: `cuvec.swig.CuVec`, raw `SWIGVector`, or `numpy.ndarray`""" + if SWIGVector.is_instance(arr): log.debug("wrap swraw %s", type(arr)) obj = np.asarray(arr).view(cls) - obj.swvec = arr + obj._vec = arr obj.cuvec = arr.cuvec return obj - if isinstance(arr, CuVec) and hasattr(arr, 'swvec'): + if isinstance(arr, CuVec) and hasattr(arr, '_vec'): log.debug("new view") obj = np.asarray(arr).view(cls) - obj.swvec = arr.swvec - obj.cuvec = arr.swvec.cuvec + obj._vec = arr._vec + obj.cuvec = arr._vec.cuvec return obj if isinstance(arr, np.ndarray): log.debug("copy") @@ -135,8 +90,8 @@ def __new__(cls, arr): raise NotImplementedError( dedent("""\ Not intended for explicit construction - (do not do `swigcuvec.CuVec((42, 1337))`; - instead use `swigcuvec.zeros((42, 137))`""")) + (do not do `cuvec.swig.CuVec((42, 1337))`; + instead use `cuvec.swig.zeros((42, 137))`""")) @property def __cuda_array_interface__(self) -> Dict[str, Any]: @@ -145,17 +100,20 @@ def __cuda_array_interface__(self) -> Dict[str, Any]: dedent("""\ `numpy.ndarray` object has no attribute `cuvec`: try using `cuvec.asarray()` first.""")) - res = self.__array_interface__ - return { - 'shape': res['shape'], 'typestr': res['typestr'], 'data': res['data'], 'version': 3} + return self._vec.__cuda_array_interface__ + + def resize(self, new_shape: Shape): + """Change shape (but not size) of array in-place.""" + self._vec.shape = new_shape + super().resize(new_shape, refcheck=False) def zeros(shape: Shape, dtype="float32") -> CuVec: """ - Returns a `swigcuvec.CuVec` view of a new `numpy.ndarray` + Returns a `cuvec.swig.CuVec` view of a new `numpy.ndarray` of the specified shape and data type (`cuvec` equivalent of `numpy.zeros`). """ - return CuVec(cu_zeros(shape, dtype)) + return CuVec(SWIGVector.zeros(shape, dtype)) ones, zeros_like, ones_like = _generate_helpers(zeros, CuVec) @@ -163,16 +121,16 @@ def zeros(shape: Shape, dtype="float32") -> CuVec: def copy(arr) -> CuVec: """ - Returns a `swigcuvec.CuVec` view of a new `numpy.ndarray` + Returns a `cuvec.swig.CuVec` view of a new `numpy.ndarray` with data copied from the specified `arr` (`cuvec` equivalent of `numpy.copy`). """ - return CuVec(cu_copy(arr)) + return CuVec(SWIGVector.copy(arr)) def asarray(arr, dtype=None, order=None, ownership: str = 'warning') -> CuVec: """ - Returns a `swigcuvec.CuVec` view of `arr`, avoiding memory copies if possible. + Returns a `cuvec.swig.CuVec` view of `arr`, avoiding memory copies if possible. (`cuvec` equivalent of `numpy.asarray`). Args: @@ -187,13 +145,13 @@ def asarray(arr, dtype=None, order=None, ownership: str = 'warning') -> CuVec: NB: `asarray()`/`retarray()` are safe if the raw cuvec was created in C++/SWIG, e.g.: >>> res = retarray(some_swig_api_func(..., output=None)) """ - if is_raw_cuvec(arr): + if SWIGVector.is_raw_cuvec(arr): ownership = ownership.lower() if ownership in {'critical', 'fatal', 'error'}: raise IOError("Can't take ownership of existing cuvec (would create dangling ptr)") getattr(log, ownership)("taking ownership") - arr = SWIGVector(None, None, arr) - if not isinstance(arr, np.ndarray) and is_raw_swvec(arr): + arr = SWIGVector('', (), arr) + if not isinstance(arr, np.ndarray) and SWIGVector.is_instance(arr): res = CuVec(arr) if dtype is None or res.dtype == np.dtype(dtype): return CuVec(np.asanyarray(res, order=order)) diff --git a/docs/index.md b/docs/index.md index 0a754fb..46e1665 100644 --- a/docs/index.md +++ b/docs/index.md @@ -3,7 +3,7 @@ Unifying Python/C++/CUDA memory: Python buffered array ↔ C++11 `std::vector` ↔ CUDA managed memory. [![Version](https://img.shields.io/pypi/v/cuvec.svg?logo=python&logoColor=white)](https://github.com/AMYPAD/CuVec/releases) -[![Downloads](https://img.shields.io/pypi/dm/cuvec.svg?logo=pypi&logoColor=white&label=PyPI%20downloads)](https://pypi.org/project/cuvec) +[![Downloads](https://img.shields.io/pypi/dm/cuvec?logo=pypi&logoColor=white)](https://pypi.org/project/cuvec) [![Py-Versions](https://img.shields.io/pypi/pyversions/cuvec.svg?logo=python&logoColor=white)](https://pypi.org/project/cuvec) [![DOI](https://zenodo.org/badge/DOI/10.5281/zenodo.4446211.svg)](https://doi.org/10.5281/zenodo.4446211) [![Licence](https://img.shields.io/pypi/l/cuvec.svg?label=licence)](https://github.com/AMYPAD/CuVec/blob/main/LICENCE) @@ -52,23 +52,25 @@ Requirements: === "Python" ```py - import cuvec - # from cuvec import swigcuvec as cuvec # SWIG alternative + import cuvec.cpython as cuvec + # import cuvec.pybind11 as cuvec # pybind11 alternative + # import cuvec.swig as cuvec # SWIG alternative arr = cuvec.zeros((1337, 42), "float32") # like `numpy.ndarray` # print(sum(arr)) # some_numpy_func(arr) # some_cpython_api_func(arr.cuvec) + # some_pybind11_func(arr.cuvec) # import cupy; cupy_arr = cupy.asarray(arr) ``` === "CPython API" ```cpp #include "Python.h" - #include "pycuvec.cuh" + #include "cuvec_cpython.cuh" PyObject *obj = (PyObject *)PyCuVec_zeros({1337, 42}); // don't forget to Py_DECREF(obj) if not returning it. - /// N.B.: convenience functions provided by "pycuvec.cuh": + /// N.B.: convenience functions provided by "cuvec_cpython.cuh": // PyCuVec *PyCuVec_zeros(std::vector shape); // PyCuVec *PyCuVec_zeros_like(PyCuVec *other); // PyCuVec *PyCuVec_deepcopy(PyCuVec *other); @@ -92,6 +94,18 @@ Requirements: // int asPyCuVec_d(PyObject *o, PyCuVec **self); ``` +=== "C++/CUDA" + ```cpp + #include "cuvec.cuh" + CuVec vec(1337 * 42); // like std::vector + ``` + +=== "C++/pybind11 API" + ```cpp + #include "cuvec.cuh" + NDCuVec ndv({1337, 42}); + ``` + === "C++/SWIG API" ```cpp #include "cuvec.cuh" @@ -105,26 +119,20 @@ Requirements: // std::vector SwigCuVec_shape(SwigCuVec *swv); ``` -=== "C++/CUDA" - ```cpp - #include "cuvec.cuh" - CuVec vec(1337 * 42); // like std::vector - ``` - ### Converting The following involve no memory copies. === "**Python** to **CPython API**" ```py - # import cuvec, my_custom_lib + # import cuvec.cpython as cuvec, my_custom_lib # arr = cuvec.zeros((1337, 42), "float32") my_custom_lib.some_cpython_api_func(arr) ``` === "**CPython API** to **Python**" ```py - import cuvec, my_custom_lib + import cuvec.cpython as cuvec, my_custom_lib arr = cuvec.asarray(my_custom_lib.some_cpython_api_func()) ``` @@ -143,17 +151,38 @@ The following involve no memory copies. float *arr = vec.data(); // pointer to `cudaMallocManaged()` data ``` +=== "**Python** to **pybind11 API**" + ```py + # import cuvec.pybind11 as cuvec, my_custom_lib + # arr = cuvec.zeros((1337, 42), "float32") + my_custom_lib.some_pybind11_api_func(arr.cuvec) + ``` + +=== "**pybind11 API** to **Python**" + ```py + import cuvec.pybind11 as cuvec, my_custom_lib + arr = cuvec.retarray(my_custom_lib.some_pybind11_api_func()) + ``` + +=== "**pybind11 API** to **C++**" + ```cpp + /// input: `NDCuVec *ndv` + /// output: `CuVec vec`, `std::vector shape` + CuVec &vec = ndv->vec; // like std::vector + std::vector &shape = ndv->shape; + ``` + === "**Python** to **SWIG API**" ```py - # import cuvec, my_custom_lib - # arr = cuvec.swigcuvec.zeros((1337, 42), "float32") + # import cuvec.swig as cuvec, my_custom_lib + # arr = cuvec.zeros((1337, 42), "float32") my_custom_lib.some_swig_api_func(arr.cuvec) ``` === "**SWIG API** to **Python**" ```py - import cuvec, my_custom_lib - arr = cuvec.swigcuvec.retarray(my_custom_lib.some_swig_api_func()) + import cuvec.swig as cuvec, my_custom_lib + arr = cuvec.retarray(my_custom_lib.some_swig_api_func()) ``` === "**SWIG API** to **C++**" @@ -180,15 +209,23 @@ Python: === "After: with CuVec" ```{.py linenums="1"} - import cuvec, numpy, mymod + import cuvec.cpython as cuvec, numpy, mymod arr = cuvec.zeros((1337, 42, 7), "float32") assert all(numpy.mean(arr, axis=(0, 1)) == 0) print(cuvec.asarray(mymod.myfunc(arr)).sum()) ``` +=== "Alternative: with CuVec & pybind11" + ```{.py linenums="1"} + import cuvec.pybind11 as cuvec, numpy, mymod + arr = cuvec.zeros((1337, 42, 7), "float32") + assert all(numpy.mean(arr, axis=(0, 1)) == 0) + print(cuvec.retarray(mymod.myfunc(arr.cuvec)).sum()) + ``` + === "Alternative: with CuVec & SWIG" ```{.py linenums="1"} - import cuvec.swigcuvec as cuvec, numpy, mymod + import cuvec.swig as cuvec, numpy, mymod arr = cuvec.zeros((1337, 42, 7), "float32") assert all(numpy.mean(arr, axis=(0, 1)) == 0) print(cuvec.retarray(mymod.myfunc(arr.cuvec)).sum()) @@ -244,7 +281,7 @@ C++: === "After: with CuVec" ```{.cpp linenums="1"} - #include "pycuvec.cuh" + #include "cuvec_cpython.cuh" #include "mycudafunction.h" static PyObject *myfunc(PyObject *self, PyObject *args, PyObject *kwargs) { @@ -262,9 +299,9 @@ C++: // hardcode upsampling factor 2 std::vector dst_shape = src->shape(); - dst_shape[2] *= 2; - dst_shape[1] *= 2; - dst_shape[0] *= 2; + for (auto &i : dst_shape) i *= 2; + + if (!dst) @@ -288,6 +325,52 @@ C++: } ``` +=== "Alternative: with CuVec & pybind11" + ```{.cpp linenums="1"} + #include "cuvec_pybind11.cuh" + #include "mycudafunction.h" + + NDCuVec *myfunc(NDCuVec &src, NDCuVec *output = nullptr) { + + + + + + + + + + + + // hardcode upsampling factor 2 + + std::vector dst_shape = src.shape; + for (auto &i : dst_shape) i *= 2; + + + + + if (!dst) + dst = new NDCuVec(dst_shape); + + + + if (!dst) throw pybind11::value_error("could not allocate output"); + + + mycudafunction(dst->vec.data(), src->vec.data(), dst_shape.data()); + + + return dst; + } + using namespace pybind11::literals; + PYBIND11_MODULE(mymod, m){ + ... + m.def("myfunc", &myfunc, "src"_a, "output"_a = nullptr); + ... + } + ``` + === "Alternative: with CuVec & SWIG" ```{.cpp linenums="1"} /// SWIG interface file mymod.i (not mymod.cpp) @@ -361,6 +444,22 @@ CUDA: + mykernel<<<...>>>(dst, src, shape[0], shape[1], shape[2]); + cudaDeviceSynchronize(); + + + } + ``` + +=== "Alternative: with CuVec & pybind11" + ```{.cpp linenums="1"} + void mycudafunction(float *dst, float *src, size_t *shape) { + + + + + + mykernel<<<...>>>(dst, src, shape[0], shape[1], shape[2]); cudaDeviceSynchronize(); @@ -384,7 +483,11 @@ CUDA: } ``` -For a full reference, see `cuvec.example_mod`'s source code: [example_mod.cu](https://github.com/AMYPAD/CuVec/blob/main/cuvec/src/example_mod/example_mod.cu) or the alternative `cuvec.example_swig` sources [example_swig.i](https://github.com/AMYPAD/CuVec/blob/main/cuvec/src/example_swig/example_swig.i) & [example_swig.cu](https://github.com/AMYPAD/CuVec/blob/main/cuvec/src/example_swig/example_swig.cu). +For a full reference, see: + +- `cuvec.example_cpython` source: [example_cpython.cu](https://github.com/AMYPAD/CuVec/blob/main/cuvec/src/example_cpython/example_cpython.cu) +- `cuvec.example_pybind11` source: [example_pybind11.cu](https://github.com/AMYPAD/CuVec/blob/main/cuvec/src/example_pybind11/example_pybind11.cu) +- `cuvec.example_swig` sources: [example_swig.i](https://github.com/AMYPAD/CuVec/blob/main/cuvec/src/example_swig/example_swig.i) & [example_swig.cu](https://github.com/AMYPAD/CuVec/blob/main/cuvec/src/example_swig/example_swig.cu) See also [NumCu](https://github.com/AMYPAD/NumCu), a minimal stand-alone Python package built using CuVec. @@ -395,14 +498,25 @@ See also [NumCu](https://github.com/AMYPAD/NumCu), a minimal stand-alone Python When using the SWIG alternative module, `arr.cuvec` is a wrapper around `SwigCuVec *`. -=== "C++ & CUDA" - `cuvec` is a header-only library so simply `#include "pycuvec.cuh"` (or `#include "cuvec.cuh"`). You can find the location of the headers using: +=== "C++/pybind11/CUDA" + `cuvec` is a header-only library so simply do one of: + + ```cpp + #include "cuvec_cpython.cuh" // CPython API + #include "cuvec_pybind11.cuh" // pybind11 API + #include "cuvec.cuh" // C++/CUDA API + ``` + + You can find the location of the headers using: ```py python -c "import cuvec; print(cuvec.include_path)" ``` - For reference, see `cuvec.example_mod`'s source code: [example_mod.cu](https://github.com/AMYPAD/CuVec/blob/main/cuvec/src/example_mod/example_mod.cu). + For reference, see: + + - `cuvec.example_cpython` source: [example_cpython.cu](https://github.com/AMYPAD/CuVec/blob/main/cuvec/src/example_cpython/example_cpython.cu) + - `cuvec.example_pybind11` source: [example_pybind11.cu](https://github.com/AMYPAD/CuVec/blob/main/cuvec/src/example_pybind11/example_pybind11.cu) === "SWIG" `cuvec` is a header-only library so simply `%include "cuvec.i"` in a SWIG interface file. You can find the location of the headers using: @@ -411,7 +525,7 @@ See also [NumCu](https://github.com/AMYPAD/NumCu), a minimal stand-alone Python python -c "import cuvec; print(cuvec.include_path)" ``` - For reference, see `cuvec.example_swig`'s source code: [example_swig.i](https://github.com/AMYPAD/CuVec/blob/main/cuvec/src/example_swig/example_swig.i) and [example_swig.cu](https://github.com/AMYPAD/CuVec/blob/main/cuvec/src/example_swig/example_swig.cu). + For reference, see `cuvec.example_swig`'s sources: [example_swig.i](https://github.com/AMYPAD/CuVec/blob/main/cuvec/src/example_swig/example_swig.i) and [example_swig.cu](https://github.com/AMYPAD/CuVec/blob/main/cuvec/src/example_swig/example_swig.cu). === "CMake" This is likely unnecessary (see the "C++ & CUDA" tab above for simpler `#include` instructions). @@ -423,14 +537,14 @@ See also [NumCu](https://github.com/AMYPAD/NumCu), a minimal stand-alone Python python -c "import cuvec; print(cuvec.cmake_prefix)" # ... or build & install directly with cmake - mkdir build && cd build - cmake ../cuvec && cmake --build . && cmake --install . --prefix /my/install/dir + cmake -S cuvec -B build && cmake --build build + cmake --install build --prefix /my/install/dir ``` At this point any external project may include `cuvec` as follows (Once setting `-DCMAKE_PREFIX_DIR=`): ```{.cmake linenums="1" hl_lines="3 6"} - cmake_minimum_required(VERSION 3.3 FATAL_ERROR) + cmake_minimum_required(VERSION 3.24 FATAL_ERROR) project(myproj) find_package(AMYPADcuvec COMPONENTS cuvec REQUIRED) add_executable(myexe ...) diff --git a/docs/pydoc-markdown.yml b/docs/pydoc-markdown.yml index a373c87..ce776aa 100644 --- a/docs/pydoc-markdown.yml +++ b/docs/pydoc-markdown.yml @@ -3,6 +3,7 @@ loaders: search_path: [..] processors: - type: filter + expression: not any(name.startswith(i) for i in ('cu_', 'is_raw', '__')) and default() - type: crossref renderer: type: mkdocs @@ -21,8 +22,17 @@ renderer: name: index source: index.md - title: Python Reference - name: ref-py - contents: [cuvec.*.CuVec, cuvec.*.zeros, cuvec.*.copy, cuvec.*.asarray] + name: ref + children: + - title: CPython + name: cpython + contents: [cuvec.cpython.*] + - title: pybind11 + name: pybind11 + contents: [cuvec.pybind11.*] + - title: SWIG + name: swig + contents: [cuvec.swig.*] - title: External Links children: - title: Source Code @@ -58,8 +68,8 @@ renderer: generator: false theme: name: material - favicon: https://github.com/squidfunk/mkdocs-material/raw/master/material/.icons/fontawesome/solid/memory.svg - logo: https://github.com/squidfunk/mkdocs-material/raw/master/material/.icons/fontawesome/solid/memory.svg + favicon: https://github.com/squidfunk/mkdocs-material/raw/master/material/templates/.icons/fontawesome/solid/memory.svg + logo: https://github.com/squidfunk/mkdocs-material/raw/master/material/templates/.icons/fontawesome/solid/memory.svg custom_dir: ../../custom_theme/ features: [content.tabs.link] palette: diff --git a/pyproject.toml b/pyproject.toml index 95f5457..3953fe2 100644 --- a/pyproject.toml +++ b/pyproject.toml @@ -1,5 +1,5 @@ [build-system] -requires = ["setuptools_scm>=7", "scikit-build-core[pyproject]>=0.5", "swig>=4"] +requires = ["setuptools_scm>=7", "scikit-build-core[pyproject]>=0.5", "swig>=4", "pybind11"] build-backend = "scikit_build_core.build" [tool.scikit-build] @@ -56,6 +56,14 @@ dependencies = ['importlib_resources; python_version < "3.9"', "numpy"] [project.optional-dependencies] dev = ["pytest>=6", "pytest-cov", "pytest-timeout", "pytest-xdist"] +[tool.mypy] +[[tool.mypy.overrides]] +module = "cuvec.swvec" +ignore_errors = true +[[tool.mypy.overrides]] +module = "cuvec.example_swig.*" +ignore_errors = true + [tool.flake8] max_line_length = 99 extend_ignore = ["E261"] @@ -84,7 +92,7 @@ timeout = 30 log_level = "INFO" python_files = ["tests/test_*.py"] testpaths = ["tests"] -addopts = "-v --tb=short -rxs -W=error -n=2 --durations=0 --durations-min=1 --cov=cuvec --cov-report=term-missing --cov-report=xml" +addopts = "-v --tb=short -rxs -W=error -n=3 --durations=0 --durations-min=1 --cov=cuvec --cov-report=term-missing --cov-report=xml" [tool.coverage.run] omit = ["cuvec/swvec.py", "cuvec/example_swig.py"] diff --git a/tests/test_common.py b/tests/test_common.py new file mode 100644 index 0000000..ec611c1 --- /dev/null +++ b/tests/test_common.py @@ -0,0 +1,209 @@ +"""Common (parameterised) tests for cuvec.{cpython,pybind11,swig}""" +import logging + +import numpy as np +from pytest import importorskip, mark, raises + +import cuvec as cu +import cuvec.cpython as cp + +try: + # `cuvec.swig` alternative to `cuvec.cpython` + # `example_swig` is defined in ../cuvec/src/example_swig/ + from cuvec import example_swig # type: ignore # yapf: disable + from cuvec import swig as sw +except ImportError: + sw, example_swig = None, None # type: ignore # yapf: disable + +try: + # `cuvec.pybind11` alternative to `cuvec.cpython` + # `example_pybind11` is defined in ../cuvec/src/example_pybind11/ + from cuvec import example_pybind11 # type: ignore # yapf: disable + from cuvec import pybind11 as py +except ImportError: + py, example_pybind11 = None, None # type: ignore # yapf: disable + + +shape = 127, 344, 344 + + +def test_includes(): + assert cu.include_path.is_dir() + assert {i.name + for i in cu.include_path.iterdir() + } == {'cuvec.cuh', 'cuvec_cpython.cuh', 'cuvec_pybind11.cuh', 'cuvec.i'} + + +def test_cmake_prefix(): + assert cu.cmake_prefix.is_dir() + assert {i.name + for i in cu.cmake_prefix.iterdir()} == { + f'AMYPADcuvec{i}.cmake' + for i in ('Config', 'ConfigVersion', 'Targets', 'Targets-release')} + + +@mark.parametrize("cu,CVector", [(py, 'Pybind11Vector'), (sw, 'SWIGVector')]) +def test_CVector_strides(cu, CVector): + v = getattr(cu, CVector)('f', shape) + a = np.asarray(v) + assert a.shape == shape + assert a.strides == (473344, 1376, 4) + + +@mark.parametrize("spec,result", [("i", np.int32), ("d", np.float64)]) +@mark.parametrize("init", ["zeros", "ones"]) +@mark.parametrize("cu", [cp, py, sw]) +def test_create(cu, init, spec, result): + a = np.asarray(getattr(cu, init)(shape, spec)) + assert a.dtype == result + assert a.shape == shape + assert (a == (0 if init == 'zeros' else 1)).all() + + b = getattr(cu, f'{init}_like')(a) + assert b.shape == a.shape + assert b.dtype == a.dtype + + +@mark.parametrize("cu", [cp, py, sw]) +def test_copy(cu): + a = np.random.random(shape) + b = np.asarray(cu.copy(a)) + assert a.shape == b.shape + assert a.dtype == b.dtype + assert (a == b).all() + + +@mark.parametrize("cu,classname", [(cp, "raw "), + (py, "pyraw "), + (sw, "swraw ")]) +def test_CuVec_creation(cu, classname, caplog): + with raises(TypeError): + cu.CuVec() + + with raises(NotImplementedError): + cu.CuVec(shape) + + caplog.set_level(logging.DEBUG) + caplog.clear() + v = cu.CuVec(np.ones(shape, dtype='h')) + assert [i[1:] for i in caplog.record_tuples] == [ + (10, 'copy'), (10, f"wrap {classname}".format(typechar='h'))] + assert v.shape == shape + assert v.dtype.char == 'h' + assert (v == 1).all() + + caplog.clear() + v = cu.zeros(shape, 'd') + assert [i[1:] for i in caplog.record_tuples] == [ + (10, f"wrap {classname}".format(typechar='d'))] + + caplog.clear() + v[0, 0, 0] = 1 + assert not caplog.record_tuples + w = cu.CuVec(v) + assert [i[1:] for i in caplog.record_tuples] == [(10, "new view")] + + caplog.clear() + assert w[0, 0, 0] == 1 + v[0, 0, 0] = 9 + assert w[0, 0, 0] == 9 + assert v.cuvec is w.cuvec + assert v.data == w.data + assert not caplog.record_tuples + + +@mark.parametrize("cu", [py, sw]) +def test_asarray(cu): + v = cu.asarray(np.random.random(shape)) + w = cu.CuVec(v) + assert w.cuvec == v.cuvec + assert (w == v).all() + assert str(w._vec) == str(v._vec) + assert np.asarray(w._vec).data == np.asarray(v._vec).data + x = cu.asarray(w._vec) + assert x.cuvec == v.cuvec + assert (x == v).all() + assert str(x._vec) == str(v._vec) + assert np.asarray(x._vec).data == np.asarray(v._vec).data + y = cu.asarray(x.tolist()) + assert y.cuvec != v.cuvec + assert (y == v).all() + assert str(y._vec) != str(v._vec) + assert np.asarray(y._vec).data == np.asarray(v._vec).data + z = cu.asarray(v[:]) + assert z.cuvec != v.cuvec + assert (z == v[:]).all() + assert str(z._vec) != str(v._vec) + assert np.asarray(z._vec).data == np.asarray(v._vec).data + s = cu.asarray(v[1:]) + assert s.cuvec != v.cuvec + assert (s == v[1:]).all() + assert str(s._vec) != str(v._vec) + assert np.asarray(s._vec).data != np.asarray(v._vec).data + with raises(IOError): + cu.asarray(s._vec.cuvec, ownership='error') + + +@mark.parametrize("cu", [py, sw]) +def test_resize(cu): + v = cu.asarray(np.random.random(shape)) + v.resize(shape[::-1]) + assert v.shape == shape[::-1] + assert v._vec.shape == v.shape + v.resize(v.size) + assert v.shape == (v.size,) + assert v._vec.shape == v.shape + + +@mark.parametrize("cu", [cp, py, sw]) +def test_cuda_array_interface(cu): + cupy = importorskip("cupy") + from cuvec import dev_sync + + v = cu.asarray(np.random.random(shape)) + assert set(v.__cuda_array_interface__) == {'shape', 'typestr', 'data', 'version'} + + c = cupy.asarray(v) + assert (c == v).all() + c[0, 0, 0] = 1 + dev_sync() + assert c[0, 0, 0] == v[0, 0, 0] + c[0, 0, 0] = 0 + dev_sync() + assert c[0, 0, 0] == v[0, 0, 0] + + if hasattr(v, '_vec'): + d = cupy.asarray(v._vec) + d[0, 0, 0] = 1 + dev_sync() + assert d[0, 0, 0] == c[0, 0, 0] == v[0, 0, 0] + d[0, 0, 0] = 0 + dev_sync() + assert d[0, 0, 0] == c[0, 0, 0] == v[0, 0, 0] + + ndarr = v + 1 + assert ndarr.shape == v.shape + assert ndarr.dtype == v.dtype + with raises(AttributeError): + ndarr.__cuda_array_interface__ + + +@mark.parametrize("cu,ex", [(py, example_pybind11), (sw, example_swig)]) +def test_increment(cu, ex): + a = cu.zeros((1337, 42), 'f') + assert (a == 0).all() + ex.increment2d_f(a.cuvec, a.cuvec) + assert (a == 1).all() + + a[:] = 0 + assert (a == 0).all() + + b = cu.retarray(ex.increment2d_f(a.cuvec)) + assert (b == 1).all() + + c = cu.retarray(ex.increment2d_f(b.cuvec, a.cuvec), a) + assert (a == 2).all() + assert c.cuvec == a.cuvec + assert (c == a).all() + assert str(c._vec) == str(a._vec) + assert np.asarray(c._vec).data == np.asarray(a._vec).data diff --git a/tests/test_cpython.py b/tests/test_cpython.py new file mode 100644 index 0000000..5a1f442 --- /dev/null +++ b/tests/test_cpython.py @@ -0,0 +1,92 @@ +import numpy as np +from pytest import mark, raises + +import cuvec.cpython as cu +from cuvec import cuvec_cpython + +shape = 127, 344, 344 + + +@mark.parametrize("tp", list(cu.typecodes)) +def test_PyCuVec_asarray(tp): + v = getattr(cuvec_cpython, f"PyCuVec_{tp}")((1, 2, 3)) + assert str(v) == f"PyCuVec_{tp}((1, 2, 3))" + a = np.asarray(v) + assert not a.any() + a[0, 0] = 42 + b = np.asarray(v) + assert (b[0, 0] == 42).all() + assert not b[1:, 1:].any() + assert a.dtype.char == tp + del a, b, v + + +def test_CVector_strides(): + v = cuvec_cpython.PyCuVec_f(shape) + a = np.asarray(v) + assert a.shape == shape + assert a.strides == (473344, 1376, 4) + + +def test_asarray(): + v = cu.asarray(np.random.random(shape)) + w = cu.CuVec(v) + assert w.cuvec == v.cuvec + assert (w == v).all() + assert np.asarray(w.cuvec).data == np.asarray(v.cuvec).data + x = cu.asarray(w.cuvec) + assert x.cuvec == v.cuvec + assert (x == v).all() + assert np.asarray(x.cuvec).data == np.asarray(v.cuvec).data + y = cu.asarray(x.tolist()) + assert y.cuvec != v.cuvec + assert (y == v).all() + assert np.asarray(y.cuvec).data == np.asarray(v.cuvec).data + z = cu.asarray(v[:]) + assert z.cuvec != v.cuvec + assert (z == v[:]).all() + assert np.asarray(z.cuvec).data == np.asarray(v.cuvec).data + s = cu.asarray(v[1:]) + assert s.cuvec != v.cuvec + assert (s == v[1:]).all() + assert np.asarray(s.cuvec).data != np.asarray(v.cuvec).data + + +def test_increment(): + # `example_cpython` is defined in ../cuvec/src/example_cpython/ + from cuvec.example_cpython import increment2d_f + a = cu.zeros((1337, 42), 'f') + assert (a == 0).all() + res = cu.asarray(increment2d_f(a.cuvec, a.cuvec)) + assert (a == 1).all() + assert (res == 1).all() + + a[:] = 0 + assert (a == 0).all() + assert (res == 0).all() + + res = cu.asarray(increment2d_f(a)) + assert (res == 1).all() + + +def test_increment_return(): + from cuvec.example_cpython import increment2d_f + a = cu.zeros((1337, 42), 'f') + assert (a == 0).all() + res = cu.asarray(increment2d_f(a, a)) + assert (a == 1).all() + del a + assert (res == 1).all() + + +def test_np_types(): + from cuvec.example_cpython import increment2d_f + f = cu.zeros((1337, 42), 'f') + d = cu.zeros((1337, 42), 'd') + cu.asarray(increment2d_f(f)) + cu.asarray(increment2d_f(f, f)) + with raises(TypeError): + cu.asarray(increment2d_f(d)) + with raises(SystemError): + # the TypeError is suppressed since a new output is generated + cu.asarray(increment2d_f(f, d)) diff --git a/tests/test_cuvec.py b/tests/test_cuvec.py deleted file mode 100644 index f036c48..0000000 --- a/tests/test_cuvec.py +++ /dev/null @@ -1,14 +0,0 @@ -import cuvec as cu - - -def test_includes(): - assert cu.include_path.is_dir() - assert {i.name for i in cu.include_path.iterdir()} == {'cuvec.cuh', 'pycuvec.cuh', 'cuvec.i'} - - -def test_cmake_prefix(): - assert cu.cmake_prefix.is_dir() - assert {i.name - for i in cu.cmake_prefix.iterdir()} == { - f'AMYPADcuvec{i}.cmake' - for i in ('Config', 'ConfigVersion', 'Targets', 'Targets-release')} diff --git a/tests/test_perf.py b/tests/test_perf.py index ef4f3e8..fd0be94 100644 --- a/tests/test_perf.py +++ b/tests/test_perf.py @@ -4,18 +4,27 @@ import numpy as np from pytest import mark, skip -# `example_mod` is defined in ../cuvec/src/example_mod/ -from cuvec import example_mod # type: ignore # yapf: disable -from cuvec import pycuvec as cu +import cuvec.cpython as cu + +# `example_cpython` is defined in ../cuvec/src/example_cpython/ +from cuvec import example_cpython # type: ignore # yapf: disable try: - # alternative to `cu` + # `cuvec.swig` alternative to `cuvec.cpython` # `example_swig` is defined in ../cuvec/src/example_swig/ from cuvec import example_swig # type: ignore # yapf: disable - from cuvec import swigcuvec as sw + from cuvec import swig as sw except ImportError: sw, example_swig = None, None # type: ignore # yapf: disable +try: + # `cuvec.pybind11` alternative to `cuvec.cpython` + # `example_pybind11` is defined in ../cuvec/src/example_pybind11/ + from cuvec import example_pybind11 # type: ignore # yapf: disable + from cuvec import pybind11 as py +except ImportError: + py, example_pybind11 = None, None # type: ignore # yapf: disable + def _time_overhead(): tic = time() @@ -52,7 +61,7 @@ def test_inner(*args, **kwargs): return wrapper -@mark.parametrize("cu,ex", [(cu, example_mod), (sw, example_swig)]) +@mark.parametrize("cu,ex", [(cu, example_cpython), (py, example_pybind11), (sw, example_swig)]) @retry_on_except() def test_perf(cu, ex, shape=(1337, 42), quiet=False, return_time=False): if cu is None: @@ -80,9 +89,9 @@ def test_perf(cu, ex, shape=(1337, 42), quiet=False, return_time=False): assert (src + 1 == dst)[1:].all() assert (src + 1 == dst)[0, 2:].all() # even a fast kernel takes longer than API overhead - assert t['- kernel'] / (t['call ext'] - t['- create dst']) > 0.5 - # API call should be <0.1 ms... but set a higher threshold of 2 ms - assert t['call ext'] - t['- create dst'] - t['- kernel'] < 2 + assert t['- kernel'] / (t['call ext'] - t['- create dst']) > 0.3 + # API call should be <0.1 ms... but set a higher threshold of 5 ms + assert t['call ext'] - t['- create dst'] - t['- kernel'] < 5 if return_time: return t @@ -94,9 +103,9 @@ def test_perf(cu, ex, shape=(1337, 42), quiet=False, return_time=False): from tqdm import trange except ImportError: trange = range - nruns = 1000 + nruns = 500 - for args in [(cu, example_mod), (sw, example_swig)]: + for args in [(cu, example_cpython), (py, example_pybind11), (sw, example_swig)]: print(f"# One run ({args[1].__name__}):") test_perf(*args, shape=(1000, 1000)) diff --git a/tests/test_pybind11.py b/tests/test_pybind11.py new file mode 100644 index 0000000..006e9f1 --- /dev/null +++ b/tests/test_pybind11.py @@ -0,0 +1,19 @@ +import numpy as np +from pytest import importorskip, mark + +cu = importorskip("cuvec.pybind11") +shape = 127, 344, 344 + + +@mark.parametrize("tp", list(cu.typecodes)) +def test_Pybind11Vector_asarray(tp): + v = cu.Pybind11Vector(tp, (1, 2, 3)) + assert repr(v) == f"Pybind11Vector('{tp}', (1, 2, 3))" + a = np.asarray(v) + assert not a.any() + a[0, 0] = 42 + b = np.asarray(v) + assert (b[0, 0] == 42).all() + assert not b[1:, 1:].any() + assert a.dtype == np.dtype(tp) + del a, b, v diff --git a/tests/test_pycuvec.py b/tests/test_pycuvec.py deleted file mode 100644 index 9f892fd..0000000 --- a/tests/test_pycuvec.py +++ /dev/null @@ -1,170 +0,0 @@ -import logging - -import numpy as np -from pytest import importorskip, mark, raises - -import cuvec as cu - -shape = 127, 344, 344 - - -@mark.parametrize("tp", list(cu.typecodes)) -def test_PyCuVec_asarray(tp): - v = getattr(cu.cuvec, f"PyCuVec_{tp}")((1, 2, 3)) - assert str(v) == f"PyCuVec_{tp}((1, 2, 3))" - a = np.asarray(v) - assert not a.any() - a[0, 0] = 42 - b = np.asarray(v) - assert (b[0, 0] == 42).all() - assert not b[1:, 1:].any() - assert a.dtype.char == tp - del a, b, v - - -def test_PyCuVec_strides(): - v = cu.cuvec.PyCuVec_f(shape) - a = np.asarray(v) - assert a.shape == shape - assert a.strides == (473344, 1376, 4) - - -@mark.parametrize("spec,result", [("i", np.int32), ("d", np.float64)]) -@mark.parametrize("init", ["zeros", "ones"]) -def test_create(init, spec, result): - a = np.asarray(getattr(cu, init)(shape, spec)) - assert a.dtype == result - assert a.shape == shape - assert (a == (0 if init == 'zeros' else 1)).all() - - b = getattr(cu, f'{init}_like')(a) - assert b.shape == a.shape - assert b.dtype == a.dtype - - -def test_copy(): - a = np.random.random(shape) - b = np.asarray(cu.copy(a)) - assert a.shape == b.shape - assert a.dtype == b.dtype - assert (a == b).all() - - -def test_CuVec_creation(caplog): - with raises(TypeError): - cu.CuVec() - - with raises(NotImplementedError): - cu.CuVec(shape) - - caplog.set_level(logging.DEBUG) - caplog.clear() - v = cu.CuVec(np.ones(shape, dtype='h')) - assert [i[1:] for i in caplog.record_tuples] == [(10, 'copy'), - (10, "wrap raw ")] - assert v.shape == shape - assert v.dtype.char == 'h' - assert (v == 1).all() - - caplog.clear() - v = cu.zeros(shape, 'd') - assert [i[1:] for i in caplog.record_tuples] == [(10, "wrap raw ")] - - caplog.clear() - v[0, 0, 0] = 1 - assert not caplog.record_tuples - w = cu.CuVec(v) - assert [i[1:] for i in caplog.record_tuples] == [(10, "new view")] - - caplog.clear() - assert w[0, 0, 0] == 1 - v[0, 0, 0] = 9 - assert w[0, 0, 0] == 9 - assert v.cuvec is w.cuvec - assert v.data == w.data - assert not caplog.record_tuples - - -def test_asarray(): - v = cu.asarray(np.random.random(shape)) - w = cu.CuVec(v) - assert w.cuvec == v.cuvec - assert (w == v).all() - assert np.asarray(w.cuvec).data == np.asarray(v.cuvec).data - x = cu.asarray(w.cuvec) - assert x.cuvec == v.cuvec - assert (x == v).all() - assert np.asarray(x.cuvec).data == np.asarray(v.cuvec).data - y = cu.asarray(x.tolist()) - assert y.cuvec != v.cuvec - assert (y == v).all() - assert np.asarray(y.cuvec).data == np.asarray(v.cuvec).data - z = cu.asarray(v[:]) - assert z.cuvec != v.cuvec - assert (z == v[:]).all() - assert np.asarray(z.cuvec).data == np.asarray(v.cuvec).data - s = cu.asarray(v[1:]) - assert s.cuvec != v.cuvec - assert (s == v[1:]).all() - assert np.asarray(s.cuvec).data != np.asarray(v.cuvec).data - - -def test_cuda_array_interface(): - cupy = importorskip("cupy") - v = cu.asarray(np.random.random(shape)) - assert hasattr(v, '__cuda_array_interface__') - - c = cupy.asarray(v) - assert (c == v).all() - c[0, 0, 0] = 1 - cu.dev_sync() - assert c[0, 0, 0] == v[0, 0, 0] - c[0, 0, 0] = 0 - cu.dev_sync() - assert c[0, 0, 0] == v[0, 0, 0] - - ndarr = v + 1 - assert ndarr.shape == v.shape - assert ndarr.dtype == v.dtype - with raises(AttributeError): - ndarr.__cuda_array_interface__ - - -def test_increment(): - # `example_mod` is defined in ../cuvec/src/example_mod/ - from cuvec.example_mod import increment2d_f - a = cu.zeros((1337, 42), 'f') - assert (a == 0).all() - res = cu.asarray(increment2d_f(a.cuvec, a.cuvec)) - assert (a == 1).all() - assert (res == 1).all() - - a[:] = 0 - assert (a == 0).all() - assert (res == 0).all() - - res = cu.asarray(increment2d_f(a)) - assert (res == 1).all() - - -def test_increment_return(): - from cuvec.example_mod import increment2d_f - a = cu.zeros((1337, 42), 'f') - assert (a == 0).all() - res = cu.asarray(increment2d_f(a, a)) - assert (a == 1).all() - del a - assert (res == 1).all() - - -def test_np_types(): - from cuvec.example_mod import increment2d_f - f = cu.zeros((1337, 42), 'f') - d = cu.zeros((1337, 42), 'd') - cu.asarray(increment2d_f(f)) - cu.asarray(increment2d_f(f, f)) - with raises(TypeError): - cu.asarray(increment2d_f(d)) - with raises(SystemError): - # the TypeError is suppressed since a new output is generated - cu.asarray(increment2d_f(f, d)) diff --git a/tests/test_swig.py b/tests/test_swig.py new file mode 100644 index 0000000..efebc34 --- /dev/null +++ b/tests/test_swig.py @@ -0,0 +1,19 @@ +import numpy as np +from pytest import importorskip, mark + +cu = importorskip("cuvec.swig") +shape = 127, 344, 344 + + +@mark.parametrize("tp", list(cu.typecodes)) +def test_SWIGVector_asarray(tp): + v = cu.SWIGVector(tp, (1, 2, 3)) + assert repr(v) == f"SWIGVector('{tp}', (1, 2, 3))" + a = np.asarray(v) + assert not a.any() + a[0, 0] = 42 + b = np.asarray(v) + assert (b[0, 0] == 42).all() + assert not b[1:, 1:].any() + assert a.dtype == np.dtype(tp) + del a, b, v diff --git a/tests/test_swigcuvec.py b/tests/test_swigcuvec.py deleted file mode 100644 index a2aee94..0000000 --- a/tests/test_swigcuvec.py +++ /dev/null @@ -1,169 +0,0 @@ -import logging - -import numpy as np -from pytest import importorskip, mark, raises - -from cuvec import dev_sync - -cu = importorskip("cuvec.swigcuvec") -shape = 127, 344, 344 - - -@mark.parametrize("tp", list(cu.typecodes)) -def test_SWIGVector_asarray(tp): - v = cu.SWIGVector(tp, (1, 2, 3)) - assert repr(v) == f"SWIGVector('{tp}', (1, 2, 3))" - a = np.asarray(v) - assert not a.any() - a[0, 0] = 42 - b = np.asarray(v) - assert (b[0, 0] == 42).all() - assert not b[1:, 1:].any() - assert a.dtype == np.dtype(tp) - del a, b, v - - -def test_PyCuVec_strides(): - v = cu.SWIGVector('f', shape) - a = np.asarray(v) - assert a.shape == shape - assert a.strides == (473344, 1376, 4) - - -@mark.parametrize("spec,result", [("i", np.int32), ("d", np.float64)]) -@mark.parametrize("init", ["zeros", "ones"]) -def test_create(init, spec, result): - a = np.asarray(getattr(cu, init)(shape, spec)) - assert a.dtype == result - assert a.shape == shape - assert (a == (0 if init == 'zeros' else 1)).all() - - b = getattr(cu, f'{init}_like')(a) - assert b.shape == a.shape - assert b.dtype == a.dtype - - -def test_copy(): - a = np.random.random(shape) - b = np.asarray(cu.copy(a)) - assert a.shape == b.shape - assert a.dtype == b.dtype - assert (a == b).all() - - -def test_CuVec_creation(caplog): - with raises(TypeError): - cu.CuVec() - - with raises(NotImplementedError): - cu.CuVec(shape) - - caplog.set_level(logging.DEBUG) - caplog.clear() - v = cu.CuVec(np.ones(shape, dtype='h')) - assert [i[1:] for i in caplog.record_tuples] == [ - (10, 'copy'), (10, "wrap swraw ")] - assert v.shape == shape - assert v.dtype.char == 'h' - assert (v == 1).all() - - caplog.clear() - v = cu.zeros(shape, 'd') - assert [i[1:] for i in caplog.record_tuples] == [ - (10, "wrap swraw ")] - - caplog.clear() - v[0, 0, 0] = 1 - assert not caplog.record_tuples - w = cu.CuVec(v) - assert [i[1:] for i in caplog.record_tuples] == [(10, "new view")] - - caplog.clear() - assert w[0, 0, 0] == 1 - v[0, 0, 0] = 9 - assert w[0, 0, 0] == 9 - assert v.cuvec is w.cuvec - assert v.data == w.data - assert not caplog.record_tuples - - -def test_asarray(): - v = cu.asarray(np.random.random(shape)) - w = cu.CuVec(v) - assert w.cuvec == v.cuvec - assert (w == v).all() - assert str(w.swvec) == str(v.swvec) - assert np.asarray(w.swvec).data == np.asarray(v.swvec).data - x = cu.asarray(w.swvec) - assert x.cuvec == v.cuvec - assert (x == v).all() - assert str(x.swvec) == str(v.swvec) - assert np.asarray(x.swvec).data == np.asarray(v.swvec).data - y = cu.asarray(x.tolist()) - assert y.cuvec != v.cuvec - assert (y == v).all() - assert str(y.swvec) != str(v.swvec) - assert np.asarray(y.swvec).data == np.asarray(v.swvec).data - z = cu.asarray(v[:]) - assert z.cuvec != v.cuvec - assert (z == v[:]).all() - assert str(z.swvec) != str(v.swvec) - assert np.asarray(z.swvec).data == np.asarray(v.swvec).data - s = cu.asarray(v[1:]) - assert s.cuvec != v.cuvec - assert (s == v[1:]).all() - assert str(s.swvec) != str(v.swvec) - assert np.asarray(s.swvec).data != np.asarray(v.swvec).data - with raises(IOError): - cu.asarray(s.swvec.cuvec, ownership='error') - - -def test_cuda_array_interface(): - cupy = importorskip("cupy") - v = cu.asarray(np.random.random(shape)) - assert hasattr(v, '__cuda_array_interface__') - - c = cupy.asarray(v) - assert (c == v).all() - c[0, 0, 0] = 1 - dev_sync() - assert c[0, 0, 0] == v[0, 0, 0] - c[0, 0, 0] = 0 - dev_sync() - assert c[0, 0, 0] == v[0, 0, 0] - - d = cupy.asarray(v.swvec) - d[0, 0, 0] = 1 - dev_sync() - assert d[0, 0, 0] == v[0, 0, 0] - d[0, 0, 0] = 0 - dev_sync() - assert d[0, 0, 0] == v[0, 0, 0] - - ndarr = v + 1 - assert ndarr.shape == v.shape - assert ndarr.dtype == v.dtype - with raises(AttributeError): - ndarr.__cuda_array_interface__ - - -def test_increment(): - # `example_swig` is defined in ../cuvec/src/example_swig/ - from cuvec.example_swig import increment2d_f - a = cu.zeros((1337, 42), 'f') - assert (a == 0).all() - increment2d_f(a.cuvec, a.cuvec) - assert (a == 1).all() - - a[:] = 0 - assert (a == 0).all() - - b = cu.retarray(increment2d_f(a.cuvec)) - assert (b == 1).all() - - c = cu.retarray(increment2d_f(b.cuvec, a.cuvec), a) - assert (a == 2).all() - assert c.cuvec == a.cuvec - assert (c == a).all() - assert str(c.swvec) == str(a.swvec) - assert np.asarray(c.swvec).data == np.asarray(a.swvec).data