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.

Sign up for GitHub

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

Jump to bottom

Forward-merge branch-24.10 into branch-24.12 #784

Merged
merged 1 commit into from
Sep 26, 2024
Merged

Forward-merge branch-24.10 into branch-24.12 #784

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
1 commit
Select commit
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
22 changes: 20 additions & 2 deletions python/cucim/src/cucim/skimage/_vendored/_ndimage_filters_core.py
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -3,12 +3,15 @@

import cupy
import numpy
from packaging.version import parse

from cucim.skimage._vendored import (
_internal as internal,
_ndimage_util as _util,
)

CUPY_GTE_13_3_0 = parse(cupy.__version__) >= parse("13.3.0")


def _origins_to_offsets(origins, w_shape):
return tuple(x // 2 + o for x, o in zip(w_shape, origins))
Expand Down Expand Up @@ -182,13 +185,24 @@ def _call_kernel(
return output


_ndimage_includes = r"""
if CUPY_GTE_13_3_0:
_includes = r"""
#include <cupy/cuda_workaround.h> // provide std:: coverage
"""
else:
_includes = r"""
#include <type_traits> // let Jitify handle this
"""

_ndimage_includes = (
_includes
+ r"""
#include <cupy/math_constants.h>

template<> struct std::is_floating_point<float16> : std::true_type {};
template<> struct std::is_signed<float16> : std::true_type {};
"""
)


_ndimage_CAST_FUNCTION = """
Expand Down Expand Up @@ -352,7 +366,11 @@ def _generate_nd_kernel(
if has_mask:
name += "_with_mask"
preamble = _ndimage_includes + _ndimage_CAST_FUNCTION + preamble
options += ("--std=c++11", "-DCUPY_USE_JITIFY")

if CUPY_GTE_13_3_0:
options += ("--std=c++11",)
else:
options += ("--std=c++11", "-DCUPY_USE_JITIFY")
return cupy.ElementwiseKernel(
in_params,
out_params,
Expand Down
6 changes: 5 additions & 1 deletion python/cucim/src/cucim/skimage/filters/_separable_filtering.py
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -5,6 +5,7 @@
from cucim.skimage._vendored import _ndimage_util as util
from cucim.skimage._vendored._internal import _normalize_axis_index
from cucim.skimage._vendored._ndimage_filters_core import (
CUPY_GTE_13_3_0,
_ndimage_CAST_FUNCTION,
_ndimage_includes,
)
Expand Down Expand Up @@ -917,7 +918,10 @@ def _get_separable_conv_kernel(
patch_per_block=patch_per_block,
flip_kernel=flip_kernel,
)
options = ("--std=c++11", "-DCUPY_USE_JITIFY")
if CUPY_GTE_13_3_0:
options = ("--std=c++11",)
else:
options = ("--std=c++11", "-DCUPY_USE_JITIFY")
m = cp.RawModule(code=code, options=options)
return m.get_function(func_name), block, patch_per_block

Expand Down
Loading