wehub-resource-sync
116 lines
5.6 KiB
C++
116 lines
5.6 KiB
C++
/* ******************************************************************************
|
|
*
|
|
*
|
|
* This program and the accompanying materials are made available under the
|
|
* terms of the Apache License, Version 2.0 which is available at
|
|
* https://www.apache.org/licenses/LICENSE-2.0.
|
|
*
|
|
* See the NOTICE file distributed with this work for additional
|
|
* information regarding copyright ownership.
|
|
* Unless required by applicable law or agreed to in writing, software
|
|
* distributed under the License is distributed on an "AS IS" BASIS, WITHOUT
|
|
* WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. See the
|
|
* License for the specific language governing permissions and limitations
|
|
* under the License.
|
|
*
|
|
* SPDX-License-Identifier: Apache-2.0
|
|
******************************************************************************/
|
|
|
|
//
|
|
// @author raver119@gmail.com
|
|
//
|
|
|
|
#ifndef PROJECT_SPECIALSSD_HOST
|
|
#define PROJECT_SPECIALSSD_HOST
|
|
#include <helpers/DebugHelper.h>
|
|
#include <helpers/shape.h>
|
|
|
|
#ifdef __CUDACC__
|
|
|
|
////////////////////////////////////////////////////////////////////////
|
|
template <typename T>
|
|
SD_HOST void bitonicSortStepGeneric(dim3 &launchDims, cudaStream_t *stream, void *vx, sd::LongType const *xShapeInfo,
|
|
int j, int k, int length, bool descending);
|
|
|
|
////////////////////////////////////////////////////////////////////////
|
|
template <typename T>
|
|
SD_HOST void bitonicArbitraryStepGeneric(dim3 &launchDims, cudaStream_t *stream, void *vx,
|
|
sd::LongType const *xShapeInfo, int window, int length, int reverse,
|
|
bool descending);
|
|
|
|
////////////////////////////////////////////////////////////////////////
|
|
template <typename X, typename Y>
|
|
SD_HOST void bitonicSortStepGenericKey(dim3 &launchDims, cudaStream_t *stream, void *vx, sd::LongType const *xShapeInfo,
|
|
void *vy, sd::LongType const *yShapeInfo, int j, int k, int length,
|
|
bool descending);
|
|
|
|
////////////////////////////////////////////////////////////////////////
|
|
template <typename X, typename Y>
|
|
SD_HOST void bitonicArbitraryStepGenericKey(dim3 &launchDims, cudaStream_t *stream, void *vx,
|
|
sd::LongType const *xShapeInfo, void *vy, sd::LongType const *yShapeInfo,
|
|
int window, int length, int reverse, bool descending);
|
|
|
|
////////////////////////////////////////////////////////////////////////
|
|
template <typename X, typename Y>
|
|
SD_HOST void bitonicSortStepGenericValue(dim3 &launchDims, cudaStream_t *stream, void *vx,
|
|
sd::LongType const *xShapeInfo, void *vy, sd::LongType const *yShapeInfo,
|
|
int j, int k, int length, bool descending);
|
|
|
|
////////////////////////////////////////////////////////////////////////
|
|
template <typename X, typename Y>
|
|
SD_HOST void bitonicArbitraryStepGenericValue(dim3 &launchDims, cudaStream_t *stream, void *vx,
|
|
sd::LongType const *xShapeInfo, void *vy, sd::LongType const *yShapeInfo,
|
|
int window, int length, int reverse, bool descending);
|
|
|
|
////////////////////////////////////////////////////////////////////////
|
|
template <typename T>
|
|
SD_HOST void oesTadGeneric(dim3 &launchDims, cudaStream_t *stream, void *vx, sd::LongType const *xShapeInfo,
|
|
sd::LongType *dimension, sd::LongType dimensionLength, sd::LongType const *tadShapeInfo,
|
|
sd::LongType const *tadOffsets, bool descending);
|
|
|
|
template <typename X, typename Y>
|
|
SD_HOST void oesTadGenericKey(dim3 &launchDims, cudaStream_t *stream, void *vx, sd::LongType const *xShapeInfo,
|
|
void *vy, sd::LongType const *yShapeInfo, sd::LongType *dimension,
|
|
sd::LongType dimensionLength,
|
|
sd::LongType const *tadShapeInfo, sd::LongType const *tadOffsets, bool descending);
|
|
|
|
template <typename X, typename Y>
|
|
SD_HOST void oesTadGenericValue(dim3 &launchDims, cudaStream_t *stream, void *vx, sd::LongType const *xShapeInfo,
|
|
void *vy, sd::LongType const *yShapeInfo, sd::LongType *dimension,
|
|
sd::LongType dimensionLength,
|
|
sd::LongType const *tadShapeInfo, sd::LongType const *tadOffsets, bool descending);
|
|
|
|
////////////////////////////////////////////////////////////////////////
|
|
template <typename T>
|
|
SD_KERNEL void printCudaGlobal(void *pointer, const int len) {
|
|
for (int i = 0; i < len; ++i) printf("%f, ", (double)reinterpret_cast<T *>(pointer)[i]);
|
|
printf("\n");
|
|
}
|
|
|
|
////////////////////////////////////////////////////////////////////////
|
|
template <typename T>
|
|
SD_DEVICE void printCudaDevice(void *pointer, const int len, const int tid = 0) {
|
|
if (blockIdx.x * blockDim.x + threadIdx.x != tid) return;
|
|
for (int i = 0; i < len; ++i) printf("%f, ", (double)reinterpret_cast<T *>(pointer)[i]);
|
|
printf("\n");
|
|
}
|
|
|
|
////////////////////////////////////////////////////////////////////////
|
|
template <typename T>
|
|
SD_HOST void printCudaHost(void *pointer, const int len, cudaStream_t &stream) {
|
|
void *ptr = malloc(sizeof(T) * len);
|
|
|
|
cudaMemcpyAsync(ptr, pointer, sizeof(T) * len, cudaMemcpyDeviceToHost, stream);
|
|
cudaError_t cudaResult = cudaStreamSynchronize(stream);
|
|
if (cudaResult != 0) THROW_EXCEPTION("printCudaHost:: cudaStreamSynchronize failed!");
|
|
|
|
for (int i = 0; i < len; ++i) printf("%f, ", (double)reinterpret_cast<T *>(ptr)[i]);
|
|
printf("\n");
|
|
|
|
free(ptr);
|
|
}
|
|
|
|
#endif
|
|
|
|
#endif // PROJECT_SPECIALSSD_HOST
|