Skip to content

Commit eb19aa3

Browse files
committed
Revert "CUDA: static assert to prevent misuse of memcpy_1 (ggml-org#17198)"
This reverts commit 5d6838b.
1 parent 905f6d9 commit eb19aa3

File tree

1 file changed

+0
-6
lines changed

1 file changed

+0
-6
lines changed

ggml/src/ggml-cuda/common.cuh

Lines changed: 0 additions & 6 deletions
Original file line numberDiff line numberDiff line change
@@ -591,12 +591,6 @@ static __device__ __forceinline__ void ggml_cuda_mad(half2 & acc, const half2 v,
591591
// If dst and src point at different address spaces then they are guaranteed to not be aliased.
592592
template <int nbytes, int alignment = 0>
593593
static __device__ __forceinline__ void ggml_cuda_memcpy_1(void * __restrict__ dst, const void * __restrict__ src) {
594-
static_assert(
595-
nbytes <= ggml_cuda_get_max_cpy_bytes() || alignment == 0,
596-
"You are misusing the alignment parameter for ggml_cuda_memcpy_1. "
597-
"The intent is for the parameter is only as a workaround if either one of the pointers is not properly aligned. "
598-
"If you use it to do more bytes per copy than ggml_cuda_max_cpy_bytes() the reads and writes may not be coalesced. "
599-
"Call ggml_cuda_memcpy_1 in a loop instead.");
600594
if constexpr (alignment != 0) {
601595
static_assert(nbytes % alignment == 0, "bad alignment");
602596
}

0 commit comments

Comments
 (0)