Merge 'master' into 'amd-master'
Change-Id: I17bb06eeadb65ee699aaf92df6d18d0677f30708
[ROCm/hip commit: f481028a31]
Цей коміт міститься в:
Jenkins
сторонній
+9
-8
@@ -151,7 +151,7 @@ def docker_build_inside_image( def build_image, String inside_args, String platf
|
||||
// The rm command needs to run as sudo because the test steps below create files owned by root
|
||||
sh """#!/usr/bin/env bash
|
||||
set -x
|
||||
sudo rm -rf ${build_dir_rel}
|
||||
rm -rf ${build_dir_rel}
|
||||
mkdir -p ${build_dir_rel}
|
||||
cd ${build_dir_rel}
|
||||
cmake -DCMAKE_BUILD_TYPE=${build_config} -DCMAKE_INSTALL_PREFIX=staging ${optional_configure} ${source_hip_abs}
|
||||
@@ -160,6 +160,7 @@ def docker_build_inside_image( def build_image, String inside_args, String platf
|
||||
}
|
||||
|
||||
// Cap the maximum amount of testing, in case of hangs
|
||||
// Excluding hipPrintfKernel test from automation; variable fails on CI test machines
|
||||
timeout(time: 1, unit: 'HOURS')
|
||||
{
|
||||
stage("${platform} unit testing")
|
||||
@@ -169,7 +170,7 @@ def docker_build_inside_image( def build_image, String inside_args, String platf
|
||||
cd ${build_dir_rel}
|
||||
make install -j\$(nproc)
|
||||
make build_tests -i -j\$(nproc)
|
||||
make test
|
||||
ctest -E hipPrintfKernel
|
||||
"""
|
||||
// If unit tests output a junit or xunit file in the future, jenkins can parse that file
|
||||
// to display test results on the dashboard
|
||||
@@ -193,7 +194,7 @@ def docker_build_inside_image( def build_image, String inside_args, String platf
|
||||
if( platform.toLowerCase( ).startsWith( 'hcc-ctu' ) )
|
||||
{
|
||||
archiveArtifacts artifacts: "${build_dir_rel}/*.deb", fingerprint: true
|
||||
archiveArtifacts artifacts: "${build_dir_rel}/*.rpm", fingerprint: true
|
||||
// archiveArtifacts artifacts: "${build_dir_rel}/*.rpm", fingerprint: true
|
||||
}
|
||||
}
|
||||
}
|
||||
@@ -367,11 +368,11 @@ if( params.hcc_integration_test )
|
||||
// The following launches 3 builds in parallel: hcc-ctu, hcc-1.6 and cuda
|
||||
parallel hcc_ctu:
|
||||
{
|
||||
node('docker && rocm')
|
||||
node('docker && rocm && dkms')
|
||||
{
|
||||
String hcc_ver = 'hcc-ctu'
|
||||
String from_image = 'compute-artifactory:5001/radeonopencompute/hcc/clang_tot_upgrade/hcc-lc-ubuntu-16.04:latest'
|
||||
String inside_args = '--device=/dev/kfd'
|
||||
String inside_args = '--device=/dev/kfd --device=/dev/dri --group-add=video'
|
||||
|
||||
// Checkout source code, dependencies and version files
|
||||
String source_hip_rel = checkout_and_version( hcc_ver )
|
||||
@@ -408,11 +409,11 @@ parallel hcc_ctu:
|
||||
},
|
||||
hcc_1_6:
|
||||
{
|
||||
node('docker && rocm')
|
||||
node('docker && rocm && !dkms')
|
||||
{
|
||||
String hcc_ver = 'hcc-1.6'
|
||||
String from_image = 'compute-artifactory:5001/radeonopencompute/hcc/roc-1.6.x/hcc-lc-ubuntu-16.04:latest'
|
||||
String inside_args = '--device=/dev/kfd'
|
||||
String from_image = 'rocm/dev-ubuntu-16.04:latest'
|
||||
String inside_args = '--device=/dev/kfd --device=/dev/dri'
|
||||
|
||||
// Checkout source code, dependencies and version files
|
||||
String source_hip_rel = checkout_and_version( hcc_ver )
|
||||
|
||||
@@ -281,7 +281,7 @@ while (@ARGV) {
|
||||
$ft{'mem'} += s/\bcudaMemcpyKind\b/hipMemcpyKind/g;
|
||||
|
||||
$ft{'mem'} += s/\bcudaPointerAttributes\b/hipPointerAttribute_t/g;
|
||||
$ft{'mem'} += s/\bcudaPointerGetAttributes\b/hipPointerGetAttribute_t/g;
|
||||
$ft{'mem'} += s/\bcudaPointerGetAttributes\b/hipPointerGetAttributes/g;
|
||||
|
||||
$ft{'mem'} += s/\bcudaMemcpy2D\b/hipMemcpy2D/g;
|
||||
$ft{'mem'} += s/\bcudaMemcpy2DToArray\b/hipMemcpy2DToArray/g;
|
||||
|
||||
@@ -7,6 +7,7 @@ MAINTAINER Kent Knox <kent.knox@amd>
|
||||
ARG user_uid
|
||||
|
||||
# Install Packages
|
||||
# python and libnuma1 are dependencies of rocm_agent_enumerator
|
||||
RUN apt-get update && DEBIAN_FRONTEND=noninteractive apt-get install -y --no-install-recommends \
|
||||
sudo \
|
||||
build-essential \
|
||||
@@ -14,6 +15,8 @@ RUN apt-get update && DEBIAN_FRONTEND=noninteractive apt-get install -y --no-ins
|
||||
git \
|
||||
libelf-dev \
|
||||
rpm \
|
||||
python \
|
||||
libnuma1 \
|
||||
&& \
|
||||
apt-get clean && \
|
||||
rm -rf /var/lib/apt/lists/*
|
||||
|
||||
@@ -5,8 +5,8 @@
|
||||
| **type** | **CUDA** | **HIP** |
|
||||
|-------------:|---------------------------------------------------------------|------------------------------------------------------------|
|
||||
| struct | `CUDA_ARRAY3D_DESCRIPTOR` | |
|
||||
| struct | `CUDA_ARRAY_DESCRIPTOR` | |
|
||||
| struct | `CUDA_MEMCPY2D` | |
|
||||
| struct | `CUDA_ARRAY_DESCRIPTOR` | `HIP_ARRAY_DESCRIPTOR` |
|
||||
| struct | `CUDA_MEMCPY2D` | `hip_Memcpy2D` |
|
||||
| struct | `CUDA_MEMCPY3D` | |
|
||||
| struct | `CUDA_MEMCPY3D_PEER` | |
|
||||
| struct | `CUDA_POINTER_ATTRIBUTE_P2P_TOKENS` | |
|
||||
@@ -27,15 +27,15 @@
|
||||
| 0x03 |*`CU_CUBEMAP_FACE_NEGATIVE_Y`* | |
|
||||
| 0x04 |*`CU_CUBEMAP_FACE_POSITIVE_Z`* | |
|
||||
| 0x05 |*`CU_CUBEMAP_FACE_NEGATIVE_Z`* | |
|
||||
| enum |***`CUarray_format`*** | |
|
||||
| 0x01 |*`CU_AD_FORMAT_UNSIGNED_INT8`* | |
|
||||
| 0x02 |*`CU_AD_FORMAT_UNSIGNED_INT16`* | |
|
||||
| 0x03 |*`CU_AD_FORMAT_UNSIGNED_INT32`* | |
|
||||
| 0x08 |*`CU_AD_FORMAT_SIGNED_INT8`* | |
|
||||
| 0x09 |*`CU_AD_FORMAT_SIGNED_INT16`* | |
|
||||
| 0x0a |*`CU_AD_FORMAT_SIGNED_INT32`* | |
|
||||
| 0x10 |*`CU_AD_FORMAT_HALF`* | |
|
||||
| 0x20 |*`CU_AD_FORMAT_FLOAT`* | |
|
||||
| enum |***`CUarray_format`*** |***`hipArray_format`*** |
|
||||
| 0x01 |*`CU_AD_FORMAT_UNSIGNED_INT8`* |*`HIP_AD_FORMAT_UNSIGNED_INT8`* |
|
||||
| 0x02 |*`CU_AD_FORMAT_UNSIGNED_INT16`* |*`HIP_AD_FORMAT_UNSIGNED_INT16`* |
|
||||
| 0x03 |*`CU_AD_FORMAT_UNSIGNED_INT32`* |*`HIP_AD_FORMAT_UNSIGNED_INT32`* |
|
||||
| 0x08 |*`CU_AD_FORMAT_SIGNED_INT8`* |*`HIP_AD_FORMAT_SIGNED_INT8`* |
|
||||
| 0x09 |*`CU_AD_FORMAT_SIGNED_INT16`* |*`HIP_AD_FORMAT_SIGNED_INT16`* |
|
||||
| 0x0a |*`CU_AD_FORMAT_SIGNED_INT32`* |*`HIP_AD_FORMAT_SIGNED_INT32`* |
|
||||
| 0x10 |*`CU_AD_FORMAT_HALF`* |*`HIP_AD_FORMAT_HALF`* |
|
||||
| 0x20 |*`CU_AD_FORMAT_FLOAT`* |*`HIP_AD_FORMAT_FLOAT`* |
|
||||
| enum |***`CUctx_flags`*** | |
|
||||
| 0x00 |*`CU_CTX_SCHED_AUTO`* | |
|
||||
| 0x01 |*`CU_CTX_SCHED_SPIN`* | |
|
||||
@@ -248,42 +248,42 @@
|
||||
| 0x01 |*`CU_RESOURCE_TYPE_MIPMAPPED_ARRAY`* | |
|
||||
| 0x02 |*`CU_RESOURCE_TYPE_LINEAR`* | |
|
||||
| 0x03 |*`CU_RESOURCE_TYPE_PITCH2D`* | |
|
||||
| enum |***`CUresourceViewFormat`*** | |
|
||||
| 0x00 |*`CU_RES_VIEW_FORMAT_NONE`* | |
|
||||
| 0x01 |*`CU_RES_VIEW_FORMAT_UINT_1X8`* | |
|
||||
| 0x02 |*`CU_RES_VIEW_FORMAT_UINT_2X8`* | |
|
||||
| 0x03 |*`CU_RES_VIEW_FORMAT_UINT_4X8`* | |
|
||||
| 0x04 |*`CU_RES_VIEW_FORMAT_SINT_1X8`* | |
|
||||
| 0x05 |*`CU_RES_VIEW_FORMAT_SINT_2X8`* | |
|
||||
| 0x06 |*`CU_RES_VIEW_FORMAT_SINT_4X8`* | |
|
||||
| 0x07 |*`CU_RES_VIEW_FORMAT_UINT_1X16`* | |
|
||||
| 0x08 |*`CU_RES_VIEW_FORMAT_UINT_2X16`* | |
|
||||
| 0x09 |*`CU_RES_VIEW_FORMAT_UINT_4X16`* | |
|
||||
| 0x0a |*`CU_RES_VIEW_FORMAT_SINT_1X16`* | |
|
||||
| 0x0b |*`CU_RES_VIEW_FORMAT_SINT_2X16`* | |
|
||||
| 0x0c |*`CU_RES_VIEW_FORMAT_SINT_4X16`* | |
|
||||
| 0x0d |*`CU_RES_VIEW_FORMAT_UINT_1X32`* | |
|
||||
| 0x0e |*`CU_RES_VIEW_FORMAT_UINT_2X32`* | |
|
||||
| 0x0f |*`CU_RES_VIEW_FORMAT_UINT_4X32`* | |
|
||||
| 0x10 |*`CU_RES_VIEW_FORMAT_SINT_1X32`* | |
|
||||
| 0x11 |*`CU_RES_VIEW_FORMAT_SINT_2X32`* | |
|
||||
| 0x12 |*`CU_RES_VIEW_FORMAT_SINT_4X32`* | |
|
||||
| 0x13 |*`CU_RES_VIEW_FORMAT_FLOAT_1X16`* | |
|
||||
| 0x14 |*`CU_RES_VIEW_FORMAT_FLOAT_2X16`* | |
|
||||
| 0x15 |*`CU_RES_VIEW_FORMAT_FLOAT_4X16`* | |
|
||||
| 0x16 |*`CU_RES_VIEW_FORMAT_FLOAT_1X32`* | |
|
||||
| 0x17 |*`CU_RES_VIEW_FORMAT_FLOAT_2X32`* | |
|
||||
| 0x18 |*`CU_RES_VIEW_FORMAT_FLOAT_4X32`* | |
|
||||
| 0x19 |*`CU_RES_VIEW_FORMAT_UNSIGNED_BC1`* | |
|
||||
| 0x1a |*`CU_RES_VIEW_FORMAT_UNSIGNED_BC3`* | |
|
||||
| 0x1b |*`CU_RES_VIEW_FORMAT_UNSIGNED_BC3`* | |
|
||||
| 0x1c |*`CU_RES_VIEW_FORMAT_UNSIGNED_BC4`* | |
|
||||
| 0x1d |*`CU_RES_VIEW_FORMAT_SIGNED_BC4`* | |
|
||||
| 0x1e |*`CU_RES_VIEW_FORMAT_UNSIGNED_BC5`* | |
|
||||
| 0x1f |*`CU_RES_VIEW_FORMAT_SIGNED_BC5`* | |
|
||||
| 0x20 |*`CU_RES_VIEW_FORMAT_UNSIGNED_BC6H`* | |
|
||||
| 0x21 |*`CU_RES_VIEW_FORMAT_SIGNED_BC6H`* | |
|
||||
| 0x22 |*`CU_RES_VIEW_FORMAT_UNSIGNED_BC7`* | |
|
||||
| enum |***`CUresourceViewFormat`*** |***`hipResourceViewFormat`*** |
|
||||
| 0x00 |*`CU_RES_VIEW_FORMAT_NONE`* |*`hipResViewFormatNone`* |
|
||||
| 0x01 |*`CU_RES_VIEW_FORMAT_UINT_1X8`* |*`hipResViewFormatUnsignedChar1`* |
|
||||
| 0x02 |*`CU_RES_VIEW_FORMAT_UINT_2X8`* |*`hipResViewFormatUnsignedChar2`* |
|
||||
| 0x03 |*`CU_RES_VIEW_FORMAT_UINT_4X8`* |*`hipResViewFormatUnsignedChar4`* |
|
||||
| 0x04 |*`CU_RES_VIEW_FORMAT_SINT_1X8`* |*`hipResViewFormatSignedChar1`* |
|
||||
| 0x05 |*`CU_RES_VIEW_FORMAT_SINT_2X8`* |*`hipResViewFormatSignedChar2`* |
|
||||
| 0x06 |*`CU_RES_VIEW_FORMAT_SINT_4X8`* |*`hipResViewFormatSignedChar4`* |
|
||||
| 0x07 |*`CU_RES_VIEW_FORMAT_UINT_1X16`* |*`hipResViewFormatUnsignedShort1`* |
|
||||
| 0x08 |*`CU_RES_VIEW_FORMAT_UINT_2X16`* |*`hipResViewFormatUnsignedShort2`* |
|
||||
| 0x09 |*`CU_RES_VIEW_FORMAT_UINT_4X16`* |*`hipResViewFormatUnsignedShort4`* |
|
||||
| 0x0a |*`CU_RES_VIEW_FORMAT_SINT_1X16`* |*`hipResViewFormatSignedShort1`* |
|
||||
| 0x0b |*`CU_RES_VIEW_FORMAT_SINT_2X16`* |*`hipResViewFormatSignedShort2`* |
|
||||
| 0x0c |*`CU_RES_VIEW_FORMAT_SINT_4X16`* |*`hipResViewFormatSignedShort4`* |
|
||||
| 0x0d |*`CU_RES_VIEW_FORMAT_UINT_1X32`* |*`hipResViewFormatUnsignedInt1`* |
|
||||
| 0x0e |*`CU_RES_VIEW_FORMAT_UINT_2X32`* |*`hipResViewFormatUnsignedInt2`* |
|
||||
| 0x0f |*`CU_RES_VIEW_FORMAT_UINT_4X32`* |*`hipResViewFormatUnsignedInt4`* |
|
||||
| 0x10 |*`CU_RES_VIEW_FORMAT_SINT_1X32`* |*`hipResViewFormatSignedInt1`* |
|
||||
| 0x11 |*`CU_RES_VIEW_FORMAT_SINT_2X32`* |*`hipResViewFormatSignedInt2`* |
|
||||
| 0x12 |*`CU_RES_VIEW_FORMAT_SINT_4X32`* |*`hipResViewFormatSignedInt4`* |
|
||||
| 0x13 |*`CU_RES_VIEW_FORMAT_FLOAT_1X16`* |*`hipResViewFormatHalf1`* |
|
||||
| 0x14 |*`CU_RES_VIEW_FORMAT_FLOAT_2X16`* |*`hipResViewFormatHalf2`* |
|
||||
| 0x15 |*`CU_RES_VIEW_FORMAT_FLOAT_4X16`* |*`hipResViewFormatHalf4`* |
|
||||
| 0x16 |*`CU_RES_VIEW_FORMAT_FLOAT_1X32`* |*`hipResViewFormatFloat1`* |
|
||||
| 0x17 |*`CU_RES_VIEW_FORMAT_FLOAT_2X32`* |*`hipResViewFormatFloat2`* |
|
||||
| 0x18 |*`CU_RES_VIEW_FORMAT_FLOAT_4X32`* |*`hipResViewFormatFloat4`* |
|
||||
| 0x19 |*`CU_RES_VIEW_FORMAT_UNSIGNED_BC1`* |*`hipResViewFormatUnsignedBlockCompressed1`* |
|
||||
| 0x1a |*`CU_RES_VIEW_FORMAT_UNSIGNED_BC2`* |*`hipResViewFormatUnsignedBlockCompressed2`* |
|
||||
| 0x1b |*`CU_RES_VIEW_FORMAT_UNSIGNED_BC3`* |*`hipResViewFormatUnsignedBlockCompressed3`* |
|
||||
| 0x1c |*`CU_RES_VIEW_FORMAT_UNSIGNED_BC4`* |*`hipResViewFormatUnsignedBlockCompressed4`* |
|
||||
| 0x1d |*`CU_RES_VIEW_FORMAT_SIGNED_BC4`* |*`hipResViewFormatSignedBlockCompressed4`* |
|
||||
| 0x1e |*`CU_RES_VIEW_FORMAT_UNSIGNED_BC5`* |*`hipResViewFormatUnsignedBlockCompressed5`* |
|
||||
| 0x1f |*`CU_RES_VIEW_FORMAT_SIGNED_BC5`* |*`hipResViewFormatSignedBlockCompressed5`* |
|
||||
| 0x20 |*`CU_RES_VIEW_FORMAT_UNSIGNED_BC6H`* |*`hipResViewFormatUnsignedBlockCompressed6H`* |
|
||||
| 0x21 |*`CU_RES_VIEW_FORMAT_SIGNED_BC6H`* |*`hipResViewFormatSignedBlockCompressed6H`* |
|
||||
| 0x22 |*`CU_RES_VIEW_FORMAT_UNSIGNED_BC7`* |*`hipResViewFormatUnsignedBlockCompressed7`* |
|
||||
| enum |***`CUresult`*** |***`hipError_t`*** |
|
||||
| 0 |*`CUDA_SUCCESS`* |*`hipSuccess`* |
|
||||
| 1 |*`CUDA_ERROR_INVALID_VALUE`* |*`hipErrorInvalidValue`* |
|
||||
@@ -394,8 +394,8 @@
|
||||
| typedef | `CUstreamCallback` | `hipStreamCallback_t` |
|
||||
| typedef | `CUsurfObject` | |
|
||||
| typedef | `CUsurfref` | |
|
||||
| typedef | `CUtexObject` | |
|
||||
| typedef | `CUtexref` | |
|
||||
| typedef | `CUtexObject` | `hipTextureObject_t` |
|
||||
| typedef | `CUtexref` | `textureReference` |
|
||||
| define |`CU_IPC_HANDLE_SIZE` | |
|
||||
| define |`CU_LAUNCH_PARAM_BUFFER_POINTER` | `HIP_LAUNCH_PARAM_BUFFER_POINTER` |
|
||||
| define |`CU_LAUNCH_PARAM_BUFFER_SIZE` | `HIP_LAUNCH_PARAM_BUFFER_SIZE` |
|
||||
@@ -507,7 +507,7 @@
|
||||
| `cuModuleGetFunction` | `hipModuleGetFunction` |
|
||||
| `cuModuleGetGlobal` | `hipModuleGetGlobal` |
|
||||
| `cuModuleGetSurfRef` | |
|
||||
| `cuModuleGetTexRef` | |
|
||||
| `cuModuleGetTexRef` | `hipModuleGetTexRef` |
|
||||
| `cuModuleLoad` | `hipModuleLoad` |
|
||||
| `cuModuleLoadData` | `hipModuleLoadData` |
|
||||
| `cuModuleLoadDataEx` | `hipModuleLoadDataEx` |
|
||||
@@ -518,7 +518,7 @@
|
||||
|
||||
| **CUDA** | **HIP** |
|
||||
|-----------------------------------------------------------|-------------------------------|
|
||||
| `cuArray3DCreate` | |
|
||||
| `cuArray3DCreate` | `hipArray3DCreate` |
|
||||
| `cuArray3DGetDescriptor` | |
|
||||
| `cuArrayCreate` | |
|
||||
| `cuArrayDestroy` | |
|
||||
@@ -674,12 +674,12 @@
|
||||
| `cuTexRefGetMipmappedArray` | |
|
||||
| `cuTexRefSetAddress` | |
|
||||
| `cuTexRefSetAddress2D` | |
|
||||
| `cuTexRefSetAddressMode` | |
|
||||
| `cuTexRefSetArray` | |
|
||||
| `cuTexRefSetAddressMode` | `hipTexRefSetAddressMode` |
|
||||
| `cuTexRefSetArray` | `hipTexRefSetArray` |
|
||||
| `cuTexRefSetBorderColor` | |
|
||||
| `cuTexRefSetFilterMode` | |
|
||||
| `cuTexRefSetFlags` | |
|
||||
| `cuTexRefSetFormat` | |
|
||||
| `cuTexRefSetFilterMode` | `hipTexRefSetFilterMode` |
|
||||
| `cuTexRefSetFlags` | `hipTexRefSetFlags` |
|
||||
| `cuTexRefSetFormat` | `hipTexRefSetFormat` |
|
||||
| `cuTexRefSetMaxAnisotropy` | |
|
||||
| `cuTexRefSetMipmapFilterMode` | |
|
||||
| `cuTexRefSetMipmapLevelBias` | |
|
||||
|
||||
@@ -124,7 +124,7 @@
|
||||
| `cudaHostRegister` | `hipHostRegister` |
|
||||
| `cudaHostUnregister` | `hipHostUnregister` |
|
||||
| `cudaMalloc` | `hipMalloc` |
|
||||
| `cudaMalloc3D` | |
|
||||
| `cudaMalloc3D` | `hipMalloc3D` |
|
||||
| `cudaMalloc3DArray` | `hipMalloc3DArray` |
|
||||
| `cudaMallocArray` | `hipMallocArray` |
|
||||
| `cudaMallocHost` | `hipHostMalloc` |
|
||||
|
||||
@@ -0,0 +1,167 @@
|
||||
# CURAND API supported by HIP
|
||||
|
||||
## **1. CURAND Data types**
|
||||
|
||||
| **type** | **CUDA** | **HIP** | **HIP value** (if differs) |
|
||||
|-------------:|---------------------------------------------------------------|------------------------------------------------------------|----------------------------|
|
||||
| enum |***`curandStatus`*** |***`hiprandStatus`*** |
|
||||
| enum |***`curandStatus_t`*** |***`hiprandStatus_t`*** |
|
||||
| 0 |*`CURAND_STATUS_SUCCESS`* |*`HIPRAND_STATUS_SUCCESS`* |
|
||||
| 100 |*`CURAND_STATUS_VERSION_MISMATCH`* |*`HIPRAND_STATUS_VERSION_MISMATCH`* |
|
||||
| 101 |*`CURAND_STATUS_NOT_INITIALIZED`* |*`HIPRAND_STATUS_NOT_INITIALIZED`* |
|
||||
| 102 |*`CURAND_STATUS_ALLOCATION_FAILED`* |*`HIPRAND_STATUS_ALLOCATION_FAILED`* |
|
||||
| 103 |*`CURAND_STATUS_TYPE_ERROR`* |*`HIPRAND_STATUS_TYPE_ERROR`* |
|
||||
| 104 |*`CURAND_STATUS_OUT_OF_RANGE`* |*`HIPRAND_STATUS_OUT_OF_RANGE`* |
|
||||
| 105 |*`CURAND_STATUS_LENGTH_NOT_MULTIPLE`* |*`HIPRAND_STATUS_LENGTH_NOT_MULTIPLE`* |
|
||||
| 106 |*`CURAND_STATUS_DOUBLE_PRECISION_REQUIRED`* |*`HIPRAND_STATUS_DOUBLE_PRECISION_REQUIRED`* |
|
||||
| 201 |*`CURAND_STATUS_LAUNCH_FAILURE`* |*`HIPRAND_STATUS_LAUNCH_FAILURE`* |
|
||||
| 202 |*`CURAND_STATUS_PREEXISTING_FAILURE`* |*`HIPRAND_STATUS_PREEXISTING_FAILURE`* |
|
||||
| 203 |*`CURAND_STATUS_INITIALIZATION_FAILED`* |*`HIPRAND_STATUS_INITIALIZATION_FAILED`* |
|
||||
| 204 |*`CURAND_STATUS_ARCH_MISMATCH`* |*`HIPRAND_STATUS_ARCH_MISMATCH`* |
|
||||
| 999 |*`CURAND_STATUS_INTERNAL_ERROR`* |*`HIPRAND_STATUS_INTERNAL_ERROR`* |
|
||||
| enum |***`curandRngType`*** |***`hiprandRngType`*** |
|
||||
| enum |***`curandRngType_t`*** |***`hiprandRngType_t`*** |
|
||||
| 0 |*`CURAND_RNG_TEST`* |*`HIPRAND_RNG_TEST`* |
|
||||
| 100 |*`CURAND_RNG_PSEUDO_DEFAULT`* |*`HIPRAND_RNG_PSEUDO_DEFAULT`* | 400 |
|
||||
| 101 |*`CURAND_RNG_PSEUDO_XORWOW`* |*`HIPRAND_RNG_PSEUDO_XORWOW`* | 401 |
|
||||
| 121 |*`CURAND_RNG_PSEUDO_MRG32K3A`* |*`HIPRAND_RNG_PSEUDO_MRG32K3A`* | 402 |
|
||||
| 141 |*`CURAND_RNG_PSEUDO_MTGP32`* |*`HIPRAND_RNG_PSEUDO_MTGP32`* | 403 |
|
||||
| 142 |*`CURAND_RNG_PSEUDO_MT19937`* |*`HIPRAND_RNG_PSEUDO_MT19937`* | 404 |
|
||||
| 161 |*`CURAND_RNG_PSEUDO_PHILOX4_32_10`* |*`HIPRAND_RNG_PSEUDO_PHILOX4_32_10`* | 405 |
|
||||
| 200 |*`CURAND_RNG_QUASI_DEFAULT`* |*`HIPRAND_RNG_QUASI_DEFAULT`* | 500 |
|
||||
| 201 |*`CURAND_RNG_QUASI_SOBOL32`* |*`HIPRAND_RNG_QUASI_SOBOL32`* | 501 |
|
||||
| 202 |*`CURAND_RNG_QUASI_SCRAMBLED_SOBOL32`* |*`HIPRAND_RNG_QUASI_SCRAMBLED_SOBOL32`* | 502 |
|
||||
| 203 |*`CURAND_RNG_QUASI_SOBOL64`* |*`HIPRAND_RNG_QUASI_SOBOL64`* | 503 |
|
||||
| 204 |*`CURAND_RNG_QUASI_SCRAMBLED_SOBOL64`* |*`HIPRAND_RNG_QUASI_SCRAMBLED_SOBOL64`* | 504 |
|
||||
| enum |***`curandOrdering`*** | |
|
||||
| enum |***`curandOrdering_t`*** | |
|
||||
| 100 |*`CURAND_ORDERING_PSEUDO_BEST`* | |
|
||||
| 101 |*`CURAND_ORDERING_PSEUDO_DEFAULT`* | |
|
||||
| 102 |*`CURAND_ORDERING_PSEUDO_SEEDED`* | |
|
||||
| 201 |*`CURAND_ORDERING_QUASI_DEFAULT`* | |
|
||||
| enum |***`curandDirectionVectorSet`*** | |
|
||||
| enum |***`curandDirectionVectorSet_t`*** | |
|
||||
| 101 |*`CURAND_DIRECTION_VECTORS_32_JOEKUO6`* | |
|
||||
| 102 |*`CURAND_SCRAMBLED_DIRECTION_VECTORS_32_JOEKUO6`* | |
|
||||
| 103 |*`CURAND_DIRECTION_VECTORS_64_JOEKUO6`* | |
|
||||
| 104 |*`CURAND_SCRAMBLED_DIRECTION_VECTORS_64_JOEKUO6`* | |
|
||||
| uint | `curandDirectionVectors32_t` | `hiprandDirectionVectors32_t` |
|
||||
| uint | `curandDirectionVectors64_t` | |
|
||||
| struct | `curandGenerator_st` | `hiprandGenerator_st` |
|
||||
| struct* | `curandGenerator_t` | `hiprandGenerator_t` |
|
||||
| enum |***`curandOrdering`*** | |
|
||||
| enum |***`curandOrdering_t`*** | |
|
||||
| 100 |*`CURAND_ORDERING_PSEUDO_BEST`* | |
|
||||
| 101 |*`CURAND_ORDERING_PSEUDO_DEFAULT`* | |
|
||||
| 102 |*`CURAND_ORDERING_PSEUDO_SEEDED`* | |
|
||||
| 201 |*`CURAND_ORDERING_QUASI_DEFAULT`* | |
|
||||
| enum |***`curandDirectionVectorSet`*** | |
|
||||
| enum |***`curandDirectionVectorSet_t`*** | |
|
||||
| 101 |*`CURAND_DIRECTION_VECTORS_32_JOEKUO6`* | |
|
||||
| 102 |*`CURAND_SCRAMBLED_DIRECTION_VECTORS_32_JOEKUO6`* | |
|
||||
| 103 |*`CURAND_DIRECTION_VECTORS_64_JOEKUO6`* | |
|
||||
| 104 |*`CURAND_SCRAMBLED_DIRECTION_VECTORS_64_JOEKUO6`* | |
|
||||
| uint | `curandDirectionVectors32_t` | `hiprandDirectionVectors32_t` |
|
||||
| uint | `curandDirectionVectors64_t` | |
|
||||
| double | `curandDistribution_st` | |
|
||||
| double | `curandHistogramM2V_st` | |
|
||||
| double* | `curandDistribution_t` | |
|
||||
| double* | `curandHistogramM2V_t` | |
|
||||
| struct | `curandDistributionShift_st` | |
|
||||
| struct* | `curandDistributionShift_t` | |
|
||||
| struct | `curandDistributionM2Shift_st` | |
|
||||
| struct* | `curandDistributionM2Shift_t` | |
|
||||
| struct | `curandHistogramM2_st` | |
|
||||
| struct* | `curandHistogramM2_t` | |
|
||||
| uint | `curandHistogramM2K_st` | |
|
||||
| uint* | `curandHistogramM2K_t` | |
|
||||
| struct | `curandDiscreteDistribution_st` | `hiprandDiscreteDistribution_st` |
|
||||
| struct* | `curandDiscreteDistribution_t` | `hiprandDiscreteDistribution_t` |
|
||||
| enum |***`curandMethod`*** | |
|
||||
| enum |***`curandMethod_t`*** | |
|
||||
| 0 |*`CURAND_CHOOSE_BEST`* | |
|
||||
| 1 |*`CURAND_ITR`* | |
|
||||
| 2 |*`CURAND_KNUTH`* | |
|
||||
| 3 |*`CURAND_HITR`* | |
|
||||
| 4 |*`CURAND_M1`* | |
|
||||
| 5 |*`CURAND_M2`* | |
|
||||
| 6 |*`CURAND_BINARY_SEARCH`* | |
|
||||
| 7 |*`CURAND_DISCRETE_GAUSS`* | |
|
||||
| 8 |*`CURAND_REJECTION`* | |
|
||||
| 9 |*`CURAND_DEVICE_API`* | |
|
||||
| 10 |*`CURAND_FAST_REJECTION`* | |
|
||||
| 11 |*`CURAND_3RD`* | |
|
||||
| 12 |*`CURAND_DEFINITION`* | |
|
||||
| 13 |*`CURAND_POISSON`* | |
|
||||
| struct | `curandStateMtgp32_t` | `hiprandStateMtgp32_t` |
|
||||
| struct | `curandStateScrambledSobol64_t` | |
|
||||
| struct | `curandStateSobol64_t` | |
|
||||
| struct | `curandStateScrambledSobol32_t` | |
|
||||
| struct | `curandStateSobol32_t` | `hiprandStateSobol32_t` |
|
||||
| struct | `curandStateMRG32k3a_t` | `hiprandStateMRG32k3a_t` |
|
||||
| struct | `curandStatePhilox4_32_10_t` | `hiprandStatePhilox4_32_10_t` |
|
||||
| struct | `curandStateXORWOW_t` | `hiprandStateXORWOW_t` |
|
||||
|
||||
## **2. Host API Functions**
|
||||
|
||||
| **CUDA** | **HIP** |
|
||||
|-----------------------------------------------------------|--------------------------------------------|
|
||||
| `curandCreateGenerator` | `hiprandCreateGenerator` |
|
||||
| `curandCreateGeneratorHost` | `hiprandCreateGeneratorHost` |
|
||||
| `curandCreatePoissonDistribution` | `hiprandCreatePoissonDistribution` |
|
||||
| `curandDestroyDistribution` | `hiprandDestroyDistribution` |
|
||||
| `curandDestroyGenerator` | `hiprandDestroyGenerator` |
|
||||
| `curandGenerate` | `hiprandGenerate` |
|
||||
| `curandGenerateLogNormal` | `hiprandGenerateLogNormal` |
|
||||
| `curandGenerateLogNormalDouble` | `hiprandGenerateLogNormalDouble` |
|
||||
| `curandGenerateLongLong` | |
|
||||
| `curandGenerateNormal` | `hiprandGenerateNormal` |
|
||||
| `curandGenerateNormalDouble` | `hiprandGenerateNormalDouble` |
|
||||
| `curandGeneratePoisson` | `hiprandGeneratePoisson` |
|
||||
| `curandGenerateSeeds` | `hiprandGenerateSeeds` |
|
||||
| `curandGenerateUniform` | `hiprandGenerateUniform` |
|
||||
| `curandGenerateUniformDouble` | `hiprandGenerateUniformDouble` |
|
||||
| `curandGetDirectionVectors32` | |
|
||||
| `curandGetDirectionVectors64` | |
|
||||
| `curandGetProperty` | |
|
||||
| `curandGetScrambleConstants32` | |
|
||||
| `curandGetScrambleConstants64` | |
|
||||
| `curandGetVersion` | `hiprandGetVersion` |
|
||||
| `curandSetGeneratorOffset` | `hiprandSetGeneratorOffset` |
|
||||
| `curandSetGeneratorOrdering` | |
|
||||
| `curandSetPseudoRandomGeneratorSeed` | `hiprandSetPseudoRandomGeneratorSeed` |
|
||||
| `curandSetQuasiRandomGeneratorDimensions` | `hiprandSetQuasiRandomGeneratorDimensions` |
|
||||
| `curandSetStream` | `hiprandSetStream` |
|
||||
|
||||
## **3. Device API Functions**
|
||||
|
||||
| **CUDA** | **HIP** |
|
||||
|-----------------------------------------------------------|--------------------------------------------|
|
||||
| `curand` | `hiprand` |
|
||||
| `curand_init` | `hiprand_init` |
|
||||
| `curand_log_normal` | `hiprand_log_normal` |
|
||||
| `curand_log_normal_double` | `hiprand_log_normal_double` |
|
||||
| `curand_log_normal2` | `hiprand_log_normal2` |
|
||||
| `curand_log_normal2_double` | `hiprand_log_normal2_double` |
|
||||
| `curand_log_normal4` | `hiprand_log_normal4` |
|
||||
| `curand_log_normal4_double` | `hiprand_log_normal4_double` |
|
||||
| `curand_mtgp32_single` | |
|
||||
| `curand_mtgp32_single_specific` | |
|
||||
| `curand_mtgp32_specific` | |
|
||||
| `curand_normal` | `hiprand_normal` |
|
||||
| `curand_normal_double` | `hiprand_normal_double` |
|
||||
| `curand_normal2` | `hiprand_normal2` |
|
||||
| `curand_normal2_double` | `hiprand_normal2_double` |
|
||||
| `curand_normal4` | `hiprand_normal4` |
|
||||
| `curand_normal4_double` | `hiprand_normal4_double` |
|
||||
| `curand_uniform2_double` | `hiprand_uniform2_double` |
|
||||
| `curand_uniform4` | `hiprand_uniform4` |
|
||||
| `curand_uniform4_double` | `hiprand_uniform4_double` |
|
||||
| `curand_discrete` | `hiprand_discrete` |
|
||||
| `curand_discrete4` | `hiprand_discrete4` |
|
||||
| `curand_poisson` | `hiprand_poisson` |
|
||||
| `curand_poisson4` | `hiprand_poisson4` |
|
||||
| `curand_Philox4x32_10` | |
|
||||
| `skipahead` | `skipahead` |
|
||||
| `skipahead_sequence` | `skipahead_sequence` |
|
||||
| `skipahead_subsequence` | `skipahead_subsequence` |
|
||||
@@ -74,7 +74,7 @@ HIP provides 3 environment variables in the HIP_*_BLOCKING family. These introd
|
||||
- HIP_API_BLOCKING : Forces hipMemcpyAsync and hipMemsetAsync to be host-synchronous, meaning they will wait for the requested operation to complete before returning to the caller.
|
||||
|
||||
These options cause HCC to serialize. Useful if you have libraries or code which is calling HCC kernels directly rather than using HIP.
|
||||
- HCC_SERIALZIE_KERNELS : 0x1=pre-serialize before each kernel launch, 0x2=post-serialize after each kernel launch., 0x3= pre- and post- serialize.
|
||||
- HCC_SERIALIZE_KERNEL : 0x1=pre-serialize before each kernel launch, 0x2=post-serialize after each kernel launch., 0x3= pre- and post- serialize.
|
||||
- HCC_SERIALIZE_COPY : 0x1=pre-serialize before each async copy, 0x2=post-serialize after each async copy., 0x3= pre- and post- serialize.
|
||||
|
||||
- HSA_ENABLE_SDMA=0 : Causes host-to-device and device-to-host copies to use compute shader blit kernels rather than the dedicated DMA copy engines. Compute shader copies have low latency (typically < 5us) and can achieve approximately 80% of the bandwidth of the DMA copy engine. This flag is useful to isolate issues with the hardware copy engines.
|
||||
|
||||
@@ -65,7 +65,7 @@ set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} -DHIPIFY_CLANG_RES=\\\"${LLVM_LIBRARY_DI
|
||||
install(TARGETS hipify-clang DESTINATION bin)
|
||||
|
||||
if (HIPIFY_CLANG_TESTS)
|
||||
find_package(PythonInterp 2.7 REQUIRED EXACT)
|
||||
find_package(PythonInterp 2.7 REQUIRED)
|
||||
|
||||
function (require_program PROGRAM_NAME)
|
||||
find_program(FOUND_${PROGRAM_NAME} ${PROGRAM_NAME})
|
||||
@@ -78,11 +78,27 @@ if (HIPIFY_CLANG_TESTS)
|
||||
|
||||
require_program(lit)
|
||||
require_program(FileCheck)
|
||||
require_program(socat)
|
||||
|
||||
# Populates CUDA_TOOLKIT_ROOT_DIR, which is then applied to the lit config to give the
|
||||
# value of --cuda-path for the test runs.
|
||||
find_package(CUDA REQUIRED)
|
||||
if ((CUDA_VERSION VERSION_LESS "7.0") OR (LLVM_PACKAGE_VERSION VERSION_LESS "3.8") OR
|
||||
(CUDA_VERSION VERSION_GREATER "7.5" AND LLVM_PACKAGE_VERSION VERSION_LESS "4.0") OR
|
||||
(CUDA_VERSION VERSION_GREATER "8.0" AND LLVM_PACKAGE_VERSION VERSION_LESS "6.0") OR
|
||||
(CUDA_VERSION VERSION_GREATER "9.0" AND LLVM_PACKAGE_VERSION VERSION_LESS "7.0"))
|
||||
message(SEND_ERROR "CUDA ${CUDA_VERSION} is not supported by clang ${LLVM_PACKAGE_VERSION}.")
|
||||
if (CUDA_VERSION VERSION_LESS "7.0")
|
||||
message(STATUS "Please install CUDA 7.0 or higher.")
|
||||
elseif ((CUDA_VERSION VERSION_EQUAL "7.0") OR (CUDA_VERSION VERSION_EQUAL "7.5"))
|
||||
message(STATUS "Please install clang 3.8 or higher.")
|
||||
elseif (CUDA_VERSION VERSION_EQUAL "8.0")
|
||||
message(STATUS "Please install clang 4.0 or higher.")
|
||||
elseif (CUDA_VERSION VERSION_EQUAL "9.0")
|
||||
message(STATUS "Please install clang 6.0 or higher.")
|
||||
elseif (CUDA_VERSION VERSION_EQUAL "9.1")
|
||||
message(STATUS "Please install clang 7.0 or higher.")
|
||||
endif()
|
||||
endif()
|
||||
|
||||
configure_file(
|
||||
${CMAKE_CURRENT_LIST_DIR}/../tests/hipify-clang/lit.site.cfg.in
|
||||
|
||||
@@ -10,8 +10,8 @@ const std::map<llvm::StringRef, hipCounter> CUDA_TYPE_NAME_MAP{
|
||||
|
||||
///////////////////////////// CUDA DRIVER API /////////////////////////////
|
||||
{"CUDA_ARRAY3D_DESCRIPTOR", {"HIP_ARRAY3D_DESCRIPTOR", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}},
|
||||
{"CUDA_ARRAY_DESCRIPTOR", {"HIP_ARRAY_DESCRIPTOR", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}},
|
||||
{"CUDA_MEMCPY2D", {"HIP_MEMCPY2D", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}},
|
||||
{"CUDA_ARRAY_DESCRIPTOR", {"HIP_ARRAY_DESCRIPTOR", CONV_TYPE, API_DRIVER}},
|
||||
{"CUDA_MEMCPY2D", {"hip_Memcpy2D", CONV_TYPE, API_DRIVER}},
|
||||
{"CUDA_MEMCPY3D", {"HIP_MEMCPY3D", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}},
|
||||
{"CUDA_MEMCPY3D_PEER", {"HIP_MEMCPY3D_PEER", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}},
|
||||
{"CUDA_POINTER_ATTRIBUTE_P2P_TOKENS", {"HIP_POINTER_ATTRIBUTE_P2P_TOKENS", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}},
|
||||
@@ -23,7 +23,7 @@ const std::map<llvm::StringRef, hipCounter> CUDA_TYPE_NAME_MAP{
|
||||
|
||||
{"CUaddress_mode", {"hipAddress_mode", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}},
|
||||
{"CUarray_cubemap_face", {"hipArray_cubemap_face", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}},
|
||||
{"CUarray_format", {"hipArray_format", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}},
|
||||
{"CUarray_format", {"hipArray_format", CONV_TYPE, API_DRIVER}},
|
||||
{"CUcomputemode", {"hipComputemode", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}}, // API_RUNTIME ANALOGUE (cudaComputeMode)
|
||||
{"CUmem_advise", {"hipMemAdvise", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}}, // API_RUNTIME ANALOGUE (cudaComputeMode)
|
||||
{"CUmem_range_attribute", {"hipMemRangeAttribute", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}}, // API_RUNTIME ANALOGUE (cudaMemRangeAttribute)
|
||||
@@ -101,8 +101,8 @@ const std::map<llvm::StringRef, hipCounter> CUDA_TYPE_NAME_MAP{
|
||||
{"CUresourcetype", {"hipResourceType", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // API_Runtime ANALOGUE (cudaResourceType)
|
||||
{"CUresourcetype_enum", {"hipResourceType", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // API_Runtime ANALOGUE (cudaResourceType)
|
||||
|
||||
{"CUresourceViewFormat", {"hipResourceViewFormat", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // API_Runtime ANALOGUE (cudaResourceViewFormat)
|
||||
{"CUresourceViewFormat_enum", {"hipResourceViewFormat", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // API_Runtime ANALOGUE (cudaResourceViewFormat)
|
||||
{"CUresourceViewFormat", {"hipResourceViewFormat", CONV_TEX, API_DRIVER}}, // API_Runtime ANALOGUE (cudaResourceViewFormat)
|
||||
{"CUresourceViewFormat_enum", {"hipResourceViewFormat", CONV_TEX, API_DRIVER}}, // API_Runtime ANALOGUE (cudaResourceViewFormat)
|
||||
|
||||
{"CUsharedconfig", {"hipSharedMemConfig", CONV_TYPE, API_DRIVER}},
|
||||
{"CUsharedconfig_enum", {"hipSharedMemConfig", CONV_TYPE, API_DRIVER}},
|
||||
@@ -128,9 +128,9 @@ const std::map<llvm::StringRef, hipCounter> CUDA_TYPE_NAME_MAP{
|
||||
// typedef struct CUsurfref_st *CUsurfref;
|
||||
{"CUsurfref", {"hipSurfaceReference_t", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}},
|
||||
// {"CUsurfref_st", {"ihipSurfaceReference_t", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}},
|
||||
{"CUtexObject", {"hipTextureObject", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}},
|
||||
{"CUtexObject", {"hipTextureObject_t", CONV_TYPE, API_DRIVER}},
|
||||
// typedef struct CUtexref_st *CUtexref;
|
||||
{"CUtexref", {"hipTextureReference_t", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}},
|
||||
{"CUtexref", {"textureReference", CONV_TYPE, API_DRIVER}},
|
||||
// {"CUtexref_st", {"ihipTextureReference_t", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}},
|
||||
|
||||
// Stream Flags enum
|
||||
@@ -213,6 +213,13 @@ const std::map<llvm::StringRef, hipCounter> CUDA_TYPE_NAME_MAP{
|
||||
{"cudaMipmappedArray_t", {"hipMipmappedArray_t", CONV_MEM, API_RUNTIME}},
|
||||
{"cudaMipmappedArray_const_t", {"hipMipmappedArray_const_t", CONV_MEM, API_RUNTIME}},
|
||||
|
||||
// defines
|
||||
{"cudaArrayDefault", {"hipArrayDefault", CONV_MEM, API_RUNTIME}},
|
||||
{"cudaArrayLayered", {"hipArrayLayered", CONV_MEM, API_RUNTIME}},
|
||||
{"cudaArraySurfaceLoadStore", {"hipArraySurfaceLoadStore", CONV_MEM, API_RUNTIME}},
|
||||
{"cudaArrayCubemap", {"hipArrayCubemap", CONV_MEM, API_RUNTIME}},
|
||||
{"cudaArrayTextureGather", {"hipArrayTextureGather", CONV_MEM, API_RUNTIME}},
|
||||
|
||||
{"cudaMemoryAdvise", {"hipMemAdvise", CONV_TYPE, API_RUNTIME, HIP_UNSUPPORTED}}, // API_Driver ANALOGUE (CUmem_advise)
|
||||
{"cudaMemRangeAttribute", {"hipMemRangeAttribute", CONV_TYPE, API_RUNTIME, HIP_UNSUPPORTED}}, // API_Driver ANALOGUE (CUmem_range_attribute)
|
||||
{"cudaMemcpyKind", {"hipMemcpyKind", CONV_MEM, API_RUNTIME}},
|
||||
@@ -263,6 +270,15 @@ const std::map<llvm::StringRef, hipCounter> CUDA_TYPE_NAME_MAP{
|
||||
|
||||
{"cudaSurfaceFormatMode", {"hipSurfaceFormatMode", CONV_SURFACE, API_RUNTIME, HIP_UNSUPPORTED}},
|
||||
|
||||
// defines
|
||||
{"cudaTextureType1D", {"hipTextureType1D", CONV_TEX, API_RUNTIME}},
|
||||
{"cudaTextureType2D", {"hipTextureType2D", CONV_TEX, API_RUNTIME}},
|
||||
{"cudaTextureType3D", {"hipTextureType3D", CONV_TEX, API_RUNTIME}},
|
||||
{"cudaTextureTypeCubemap", {"hipTextureTypeCubemap", CONV_TEX, API_RUNTIME}},
|
||||
{"cudaTextureType1DLayered", {"hipTextureType1DLayered", CONV_TEX, API_RUNTIME}},
|
||||
{"cudaTextureType2DLayered", {"hipTextureType2DLayered", CONV_TEX, API_RUNTIME}},
|
||||
{"cudaTextureTypeCubemapLayered", {"hipTextureTypeCubemapLayered", CONV_TEX, API_RUNTIME}},
|
||||
|
||||
// Inter-Process Communication (IPC)
|
||||
{"cudaIpcEventHandle_t", {"hipIpcEventHandle_t", CONV_TYPE, API_RUNTIME}},
|
||||
{"cudaIpcEventHandle_st", {"hipIpcEventHandle_t", CONV_TYPE, API_RUNTIME}},
|
||||
@@ -301,12 +317,51 @@ const std::map<llvm::StringRef, hipCounter> CUDA_TYPE_NAME_MAP{
|
||||
|
||||
{"cublasOperation_t", {"hipblasOperation_t", CONV_TYPE, API_BLAS}},
|
||||
{"cublasStatus_t", {"hipblasStatus_t", CONV_TYPE, API_BLAS}},
|
||||
{"cublasFillMode_t", {"hipblasFillMode_t", CONV_TYPE, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasDiagType_t", {"hipblasDiagType_t", CONV_TYPE, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasSideMode_t", {"hipblasSideMode_t", CONV_TYPE, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasPointerMode_t", {"hipblasPointerMode_t", CONV_TYPE, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasFillMode_t", {"hipblasFillMode_t", CONV_TYPE, API_BLAS}},
|
||||
{"cublasDiagType_t", {"hipblasDiagType_t", CONV_TYPE, API_BLAS}},
|
||||
{"cublasSideMode_t", {"hipblasSideMode_t", CONV_TYPE, API_BLAS}},
|
||||
{"cublasPointerMode_t", {"hipblasPointerMode_t", CONV_TYPE, API_BLAS}},
|
||||
{"cublasAtomicsMode_t", {"hipblasAtomicsMode_t", CONV_TYPE, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasDataType_t", {"hipblasDataType_t", CONV_TYPE, API_BLAS, HIP_UNSUPPORTED}},
|
||||
|
||||
///////////////////////////// cuRAND /////////////////////////////
|
||||
{"curandStatus", {"hiprandStatus_t", CONV_TYPE, API_RAND}},
|
||||
{"curandStatus_t", {"hiprandStatus_t", CONV_TYPE, API_RAND}},
|
||||
{"curandRngType", {"hiprandRngType_t", CONV_TYPE, API_RAND}},
|
||||
{"curandRngType_t", {"hiprandRngType_t", CONV_TYPE, API_RAND}},
|
||||
{"curandGenerator_st", {"hiprandGenerator_st", CONV_TYPE, API_RAND}},
|
||||
{"curandGenerator_t", {"hiprandGenerator_t", CONV_TYPE, API_RAND}},
|
||||
{"curandDirectionVectorSet", {"hiprandDirectionVectorSet_t", CONV_TYPE, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"curandDirectionVectorSet_t", {"hiprandDirectionVectorSet_t", CONV_TYPE, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"curandOrdering", {"hiprandOrdering_t", CONV_TYPE, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"curandOrdering_t", {"hiprandOrdering_t", CONV_TYPE, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"curandDistribution_st", {"hiprandDistribution_st", CONV_TYPE, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"curandHistogramM2V_st", {"hiprandDistribution_st", CONV_TYPE, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"curandDistribution_t", {"hiprandDistribution_t", CONV_TYPE, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"curandHistogramM2V_t", {"hiprandDistribution_t", CONV_TYPE, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"curandDistributionShift_st", {"hiprandDistributionShift_st", CONV_TYPE, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"curandDistributionShift_t", {"hiprandDistributionShift_t", CONV_TYPE, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"curandDistributionM2Shift_st", {"hiprandDistributionM2Shift_st", CONV_TYPE, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"curandDistributionM2Shift_t", {"hiprandDistributionM2Shift_t", CONV_TYPE, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"curandHistogramM2_st", {"hiprandHistogramM2_st", CONV_TYPE, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"curandHistogramM2_t", {"hiprandHistogramM2_t", CONV_TYPE, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"curandHistogramM2K_st", {"hiprandHistogramM2K_st", CONV_TYPE, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"curandHistogramM2K_t", {"hiprandHistogramM2K_t", CONV_TYPE, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"curandDiscreteDistribution_st", {"hiprandDiscreteDistribution_st", CONV_TYPE, API_RAND}},
|
||||
{"curandDiscreteDistribution_t", {"hiprandDiscreteDistribution_t", CONV_TYPE, API_RAND}},
|
||||
{"curandMethod", {"hiprandMethod_t", CONV_TYPE, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"curandMethod_t", {"hiprandMethod_t", CONV_TYPE, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"curandDirectionVectors32_t", {"hiprandDirectionVectors32_t", CONV_TYPE, API_RAND}},
|
||||
{"curandDirectionVectors64_t", {"hiprandDirectionVectors64_t", CONV_TYPE, API_RAND, HIP_UNSUPPORTED}},
|
||||
// cuRAND types for Device functions
|
||||
{"curandStateMtgp32_t", {"hiprandStateMtgp32_t", CONV_TYPE, API_RAND}},
|
||||
{"curandStateScrambledSobol64_t", {"hiprandStateScrambledSobol64_t", CONV_TYPE, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"curandStateSobol64_t", {"hiprandStateSobol64_t", CONV_TYPE, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"curandStateScrambledSobol32_t", {"hiprandStateScrambledSobol32_t", CONV_TYPE, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"curandStateSobol32_t", {"hiprandStateSobol32_t", CONV_TYPE, API_RAND}},
|
||||
{"curandStateMRG32k3a_t", {"hiprandStateMRG32k3a_t", CONV_TYPE, API_RAND}},
|
||||
{"curandStatePhilox4_32_10_t", {"hiprandStatePhilox4_32_10_t", CONV_TYPE, API_RAND}},
|
||||
{"curandStateXORWOW_t", {"hiprandStateXORWOW_t", CONV_TYPE, API_RAND}},
|
||||
};
|
||||
|
||||
/// Maps cuda header names to hip header names.
|
||||
@@ -324,8 +379,12 @@ const std::map <llvm::StringRef, hipCounter> CUDA_INCLUDE_MAP{
|
||||
{"vector_types.h", {"hip/hip_vector_types.h", CONV_INCLUDE, API_RUNTIME}},
|
||||
|
||||
// CUBLAS includes
|
||||
{"cublas.h", {"hipblas.h", CONV_INCLUDE, API_BLAS}},
|
||||
{"cublas_v2.h", {"hipblas.h", CONV_INCLUDE, API_BLAS}},
|
||||
{"cublas.h", {"hipblas.h", CONV_INCLUDE_CUDA_MAIN_H, API_BLAS}},
|
||||
{"cublas_v2.h", {"hipblas.h", CONV_INCLUDE_CUDA_MAIN_H, API_BLAS}},
|
||||
|
||||
// CURAND includes
|
||||
{"curand.h", {"hiprand.h", CONV_INCLUDE, API_RAND}},
|
||||
{"curand_kernel.h", {"hiprand_kernel.h", CONV_INCLUDE, API_RAND}},
|
||||
|
||||
// HIP includes
|
||||
// TODO: uncomment this when hip/cudacommon.h will be renamed to hip/hipcommon.h
|
||||
@@ -548,14 +607,14 @@ const std::map<llvm::StringRef, hipCounter> CUDA_IDENTIFIER_MAP{
|
||||
{"CU_CUBEMAP_FACE_NEGATIVE_Z", {"HIP_CUBEMAP_FACE_NEGATIVE_Z", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}}, // 0x05
|
||||
|
||||
// CUarray_format enum
|
||||
{"CU_AD_FORMAT_UNSIGNED_INT8", {"HIP_AD_FORMAT_UNSIGNED_INT8", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}}, // 0x01
|
||||
{"CU_AD_FORMAT_UNSIGNED_INT16", {"HIP_AD_FORMAT_UNSIGNED_INT16", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}}, // 0x02
|
||||
{"CU_AD_FORMAT_UNSIGNED_INT32", {"HIP_AD_FORMAT_UNSIGNED_INT32", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}}, // 0x03
|
||||
{"CU_AD_FORMAT_SIGNED_INT8", {"HIP_AD_FORMAT_SIGNED_INT8", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}}, // 0x08
|
||||
{"CU_AD_FORMAT_SIGNED_INT16", {"HIP_AD_FORMAT_SIGNED_INT16", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}}, // 0x09
|
||||
{"CU_AD_FORMAT_SIGNED_INT32", {"HIP_AD_FORMAT_SIGNED_INT32", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}}, // 0x0a
|
||||
{"CU_AD_FORMAT_HALF", {"HIP_AD_FORMAT_HALF", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}}, // 0x10
|
||||
{"CU_AD_FORMAT_FLOAT", {"HIP_AD_FORMAT_FLOAT", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}}, // 0x20
|
||||
{"CU_AD_FORMAT_UNSIGNED_INT8", {"HIP_AD_FORMAT_UNSIGNED_INT8", CONV_TYPE, API_DRIVER}}, // 0x01
|
||||
{"CU_AD_FORMAT_UNSIGNED_INT16", {"HIP_AD_FORMAT_UNSIGNED_INT16", CONV_TYPE, API_DRIVER}}, // 0x02
|
||||
{"CU_AD_FORMAT_UNSIGNED_INT32", {"HIP_AD_FORMAT_UNSIGNED_INT32", CONV_TYPE, API_DRIVER}}, // 0x03
|
||||
{"CU_AD_FORMAT_SIGNED_INT8", {"HIP_AD_FORMAT_SIGNED_INT8", CONV_TYPE, API_DRIVER}}, // 0x08
|
||||
{"CU_AD_FORMAT_SIGNED_INT16", {"HIP_AD_FORMAT_SIGNED_INT16", CONV_TYPE, API_DRIVER}}, // 0x09
|
||||
{"CU_AD_FORMAT_SIGNED_INT32", {"HIP_AD_FORMAT_SIGNED_INT32", CONV_TYPE, API_DRIVER}}, // 0x0a
|
||||
{"CU_AD_FORMAT_HALF", {"HIP_AD_FORMAT_HALF", CONV_TYPE, API_DRIVER}}, // 0x10
|
||||
{"CU_AD_FORMAT_FLOAT", {"HIP_AD_FORMAT_FLOAT", CONV_TYPE, API_DRIVER}}, // 0x20
|
||||
|
||||
// CUcomputemode enum
|
||||
{"CU_COMPUTEMODE_DEFAULT", {"hipComputeModeDefault", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}}, // 0 // API_RUNTIME ANALOGUE (cudaComputeModeDefault = 0)
|
||||
@@ -855,41 +914,41 @@ const std::map<llvm::StringRef, hipCounter> CUDA_IDENTIFIER_MAP{
|
||||
{"CU_RESOURCE_TYPE_PITCH2D", {"hipResourceTypePitch2D", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x03 // API_Runtime ANALOGUE (cudaResourceTypePitch2D = 0x03)
|
||||
|
||||
// enum CUresourceViewFormat/CUresourceViewFormat_enum
|
||||
{"CU_RES_VIEW_FORMAT_NONE", {"hipResViewFormatNone", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x00 // API_Runtime ANALOGUE (cudaResViewFormatNone = 0x00)
|
||||
{"CU_RES_VIEW_FORMAT_UINT_1X8", {"hipResViewFormatUnsignedChar1", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x01 // API_Runtime ANALOGUE (cudaResViewFormatUnsignedChar1 = 0x01)
|
||||
{"CU_RES_VIEW_FORMAT_UINT_2X8", {"hipResViewFormatUnsignedChar2", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x02 // API_Runtime ANALOGUE (cudaResViewFormatUnsignedChar2 = 0x02)
|
||||
{"CU_RES_VIEW_FORMAT_UINT_4X8", {"hipResViewFormatUnsignedChar4", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x03 // API_Runtime ANALOGUE (cudaResViewFormatUnsignedChar4 = 0x03)
|
||||
{"CU_RES_VIEW_FORMAT_SINT_1X8", {"hipResViewFormatSignedChar1", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x04 // API_Runtime ANALOGUE (cudaResViewFormatSignedChar1 = 0x04)
|
||||
{"CU_RES_VIEW_FORMAT_SINT_2X8", {"hipResViewFormatSignedChar2", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x05 // API_Runtime ANALOGUE (cudaResViewFormatSignedChar2 = 0x05)
|
||||
{"CU_RES_VIEW_FORMAT_SINT_4X8", {"hipResViewFormatSignedChar4", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x06 // API_Runtime ANALOGUE (cudaResViewFormatSignedChar4 = 0x06)
|
||||
{"CU_RES_VIEW_FORMAT_UINT_1X16", {"hipResViewFormatUnsignedShort1", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x07 // API_Runtime ANALOGUE (cudaResViewFormatUnsignedShort1 = 0x07)
|
||||
{"CU_RES_VIEW_FORMAT_UINT_2X16", {"hipResViewFormatUnsignedShort2", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x08 // API_Runtime ANALOGUE (cudaResViewFormatUnsignedShort2 = 0x08)
|
||||
{"CU_RES_VIEW_FORMAT_UINT_4X16", {"hipResViewFormatUnsignedShort4", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x09 // API_Runtime ANALOGUE (cudaResViewFormatUnsignedShort4 = 0x09)
|
||||
{"CU_RES_VIEW_FORMAT_SINT_1X16", {"hipResViewFormatSignedShort1", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x0a // API_Runtime ANALOGUE (cudaResViewFormatSignedShort1 = 0x0a)
|
||||
{"CU_RES_VIEW_FORMAT_SINT_2X16", {"hipResViewFormatSignedShort2", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x0b // API_Runtime ANALOGUE (cudaResViewFormatSignedShort2 = 0x0b)
|
||||
{"CU_RES_VIEW_FORMAT_SINT_4X16", {"hipResViewFormatSignedShort4", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x0c // API_Runtime ANALOGUE (cudaResViewFormatSignedShort4 = 0x0c)
|
||||
{"CU_RES_VIEW_FORMAT_UINT_1X32", {"hipResViewFormatUnsignedInt1", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x0d // API_Runtime ANALOGUE (cudaResViewFormatUnsignedInt1 = 0x0d)
|
||||
{"CU_RES_VIEW_FORMAT_UINT_2X32", {"hipResViewFormatUnsignedInt2", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x0e // API_Runtime ANALOGUE (cudaResViewFormatUnsignedInt2 = 0x0e)
|
||||
{"CU_RES_VIEW_FORMAT_UINT_4X32", {"hipResViewFormatUnsignedInt4", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x0f // API_Runtime ANALOGUE (cudaResViewFormatUnsignedInt4 = 0x0f)
|
||||
{"CU_RES_VIEW_FORMAT_SINT_1X32", {"hipResViewFormatSignedInt1", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x10 // API_Runtime ANALOGUE (cudaResViewFormatSignedInt1 = 0x10)
|
||||
{"CU_RES_VIEW_FORMAT_SINT_2X32", {"hipResViewFormatSignedInt2", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x11 // API_Runtime ANALOGUE (cudaResViewFormatSignedInt2 = 0x11)
|
||||
{"CU_RES_VIEW_FORMAT_SINT_4X32", {"hipResViewFormatSignedInt4", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x12 // API_Runtime ANALOGUE (cudaResViewFormatSignedInt4 = 0x12)
|
||||
{"CU_RES_VIEW_FORMAT_FLOAT_1X16", {"hipResViewFormatHalf1", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x13 // API_Runtime ANALOGUE (cudaResViewFormatHalf1 = 0x13)
|
||||
{"CU_RES_VIEW_FORMAT_FLOAT_2X16", {"hipResViewFormatHalf2", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x14 // API_Runtime ANALOGUE (cudaResViewFormatHalf2 = 0x14)
|
||||
{"CU_RES_VIEW_FORMAT_FLOAT_4X16", {"hipResViewFormatHalf4", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x15 // API_Runtime ANALOGUE (cudaResViewFormatHalf4 = 0x15)
|
||||
{"CU_RES_VIEW_FORMAT_FLOAT_1X32", {"hipResViewFormatFloat1", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x16 // API_Runtime ANALOGUE (cudaResViewFormatFloat1 = 0x16)
|
||||
{"CU_RES_VIEW_FORMAT_FLOAT_2X32", {"hipResViewFormatFloat2", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x17 // API_Runtime ANALOGUE (cudaResViewFormatFloat2 = 0x17)
|
||||
{"CU_RES_VIEW_FORMAT_FLOAT_4X32", {"hipResViewFormatFloat4", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x18 // API_Runtime ANALOGUE (cudaResViewFormatFloat4 = 0x18)
|
||||
{"CU_RES_VIEW_FORMAT_UNSIGNED_BC1", {"hipResViewFormatUnsignedBlockCompressed1", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x19 // API_Runtime ANALOGUE (cudaResViewFormatUnsignedBlockCompressed1 = 0x19)
|
||||
{"CU_RES_VIEW_FORMAT_UNSIGNED_BC2", {"hipResViewFormatUnsignedBlockCompressed2", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x1a // API_Runtime ANALOGUE (cudaResViewFormatUnsignedBlockCompressed2 = 0x1a)
|
||||
{"CU_RES_VIEW_FORMAT_UNSIGNED_BC3", {"hipResViewFormatUnsignedBlockCompressed3", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x1b // API_Runtime ANALOGUE (cudaResViewFormatUnsignedBlockCompressed3 = 0x1b)
|
||||
{"CU_RES_VIEW_FORMAT_UNSIGNED_BC4", {"hipResViewFormatUnsignedBlockCompressed4", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x1c // API_Runtime ANALOGUE (cudaResViewFormatUnsignedBlockCompressed4 = 0x1c)
|
||||
{"CU_RES_VIEW_FORMAT_SIGNED_BC4", {"hipResViewFormatSignedBlockCompressed4", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x1d // API_Runtime ANALOGUE (cudaResViewFormatSignedBlockCompressed4 = 0x1d)
|
||||
{"CU_RES_VIEW_FORMAT_UNSIGNED_BC5", {"hipResViewFormatUnsignedBlockCompressed5", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x1e // API_Runtime ANALOGUE (cudaResViewFormatUnsignedBlockCompressed5 = 0x1e)
|
||||
{"CU_RES_VIEW_FORMAT_SIGNED_BC5", {"hipResViewFormatSignedBlockCompressed5", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x1f // API_Runtime ANALOGUE (cudaResViewFormatSignedBlockCompressed5 = 0x1f)
|
||||
{"CU_RES_VIEW_FORMAT_UNSIGNED_BC6H", {"hipResViewFormatUnsignedBlockCompressed6H", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x20 // API_Runtime ANALOGUE (cudaResViewFormatUnsignedBlockCompressed6H = 0x20)
|
||||
{"CU_RES_VIEW_FORMAT_SIGNED_BC6H", {"hipResViewFormatSignedBlockCompressed6H", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x21 // API_Runtime ANALOGUE (cudaResViewFormatSignedBlockCompressed6H = 0x21)
|
||||
{"CU_RES_VIEW_FORMAT_UNSIGNED_BC7", {"hipResViewFormatUnsignedBlockCompressed7", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // 0x22 // API_Runtime ANALOGUE (cudaResViewFormatUnsignedBlockCompressed7 = 0x22)
|
||||
{"CU_RES_VIEW_FORMAT_NONE", {"hipResViewFormatNone", CONV_TEX, API_DRIVER}}, // 0x00 // API_Runtime ANALOGUE (cudaResViewFormatNone = 0x00)
|
||||
{"CU_RES_VIEW_FORMAT_UINT_1X8", {"hipResViewFormatUnsignedChar1", CONV_TEX, API_DRIVER}}, // 0x01 // API_Runtime ANALOGUE (cudaResViewFormatUnsignedChar1 = 0x01)
|
||||
{"CU_RES_VIEW_FORMAT_UINT_2X8", {"hipResViewFormatUnsignedChar2", CONV_TEX, API_DRIVER}}, // 0x02 // API_Runtime ANALOGUE (cudaResViewFormatUnsignedChar2 = 0x02)
|
||||
{"CU_RES_VIEW_FORMAT_UINT_4X8", {"hipResViewFormatUnsignedChar4", CONV_TEX, API_DRIVER}}, // 0x03 // API_Runtime ANALOGUE (cudaResViewFormatUnsignedChar4 = 0x03)
|
||||
{"CU_RES_VIEW_FORMAT_SINT_1X8", {"hipResViewFormatSignedChar1", CONV_TEX, API_DRIVER}}, // 0x04 // API_Runtime ANALOGUE (cudaResViewFormatSignedChar1 = 0x04)
|
||||
{"CU_RES_VIEW_FORMAT_SINT_2X8", {"hipResViewFormatSignedChar2", CONV_TEX, API_DRIVER}}, // 0x05 // API_Runtime ANALOGUE (cudaResViewFormatSignedChar2 = 0x05)
|
||||
{"CU_RES_VIEW_FORMAT_SINT_4X8", {"hipResViewFormatSignedChar4", CONV_TEX, API_DRIVER}}, // 0x06 // API_Runtime ANALOGUE (cudaResViewFormatSignedChar4 = 0x06)
|
||||
{"CU_RES_VIEW_FORMAT_UINT_1X16", {"hipResViewFormatUnsignedShort1", CONV_TEX, API_DRIVER}}, // 0x07 // API_Runtime ANALOGUE (cudaResViewFormatUnsignedShort1 = 0x07)
|
||||
{"CU_RES_VIEW_FORMAT_UINT_2X16", {"hipResViewFormatUnsignedShort2", CONV_TEX, API_DRIVER}}, // 0x08 // API_Runtime ANALOGUE (cudaResViewFormatUnsignedShort2 = 0x08)
|
||||
{"CU_RES_VIEW_FORMAT_UINT_4X16", {"hipResViewFormatUnsignedShort4", CONV_TEX, API_DRIVER}}, // 0x09 // API_Runtime ANALOGUE (cudaResViewFormatUnsignedShort4 = 0x09)
|
||||
{"CU_RES_VIEW_FORMAT_SINT_1X16", {"hipResViewFormatSignedShort1", CONV_TEX, API_DRIVER}}, // 0x0a // API_Runtime ANALOGUE (cudaResViewFormatSignedShort1 = 0x0a)
|
||||
{"CU_RES_VIEW_FORMAT_SINT_2X16", {"hipResViewFormatSignedShort2", CONV_TEX, API_DRIVER}}, // 0x0b // API_Runtime ANALOGUE (cudaResViewFormatSignedShort2 = 0x0b)
|
||||
{"CU_RES_VIEW_FORMAT_SINT_4X16", {"hipResViewFormatSignedShort4", CONV_TEX, API_DRIVER}}, // 0x0c // API_Runtime ANALOGUE (cudaResViewFormatSignedShort4 = 0x0c)
|
||||
{"CU_RES_VIEW_FORMAT_UINT_1X32", {"hipResViewFormatUnsignedInt1", CONV_TEX, API_DRIVER}}, // 0x0d // API_Runtime ANALOGUE (cudaResViewFormatUnsignedInt1 = 0x0d)
|
||||
{"CU_RES_VIEW_FORMAT_UINT_2X32", {"hipResViewFormatUnsignedInt2", CONV_TEX, API_DRIVER}}, // 0x0e // API_Runtime ANALOGUE (cudaResViewFormatUnsignedInt2 = 0x0e)
|
||||
{"CU_RES_VIEW_FORMAT_UINT_4X32", {"hipResViewFormatUnsignedInt4", CONV_TEX, API_DRIVER}}, // 0x0f // API_Runtime ANALOGUE (cudaResViewFormatUnsignedInt4 = 0x0f)
|
||||
{"CU_RES_VIEW_FORMAT_SINT_1X32", {"hipResViewFormatSignedInt1", CONV_TEX, API_DRIVER}}, // 0x10 // API_Runtime ANALOGUE (cudaResViewFormatSignedInt1 = 0x10)
|
||||
{"CU_RES_VIEW_FORMAT_SINT_2X32", {"hipResViewFormatSignedInt2", CONV_TEX, API_DRIVER}}, // 0x11 // API_Runtime ANALOGUE (cudaResViewFormatSignedInt2 = 0x11)
|
||||
{"CU_RES_VIEW_FORMAT_SINT_4X32", {"hipResViewFormatSignedInt4", CONV_TEX, API_DRIVER}}, // 0x12 // API_Runtime ANALOGUE (cudaResViewFormatSignedInt4 = 0x12)
|
||||
{"CU_RES_VIEW_FORMAT_FLOAT_1X16", {"hipResViewFormatHalf1", CONV_TEX, API_DRIVER}}, // 0x13 // API_Runtime ANALOGUE (cudaResViewFormatHalf1 = 0x13)
|
||||
{"CU_RES_VIEW_FORMAT_FLOAT_2X16", {"hipResViewFormatHalf2", CONV_TEX, API_DRIVER}}, // 0x14 // API_Runtime ANALOGUE (cudaResViewFormatHalf2 = 0x14)
|
||||
{"CU_RES_VIEW_FORMAT_FLOAT_4X16", {"hipResViewFormatHalf4", CONV_TEX, API_DRIVER}}, // 0x15 // API_Runtime ANALOGUE (cudaResViewFormatHalf4 = 0x15)
|
||||
{"CU_RES_VIEW_FORMAT_FLOAT_1X32", {"hipResViewFormatFloat1", CONV_TEX, API_DRIVER}}, // 0x16 // API_Runtime ANALOGUE (cudaResViewFormatFloat1 = 0x16)
|
||||
{"CU_RES_VIEW_FORMAT_FLOAT_2X32", {"hipResViewFormatFloat2", CONV_TEX, API_DRIVER}}, // 0x17 // API_Runtime ANALOGUE (cudaResViewFormatFloat2 = 0x17)
|
||||
{"CU_RES_VIEW_FORMAT_FLOAT_4X32", {"hipResViewFormatFloat4", CONV_TEX, API_DRIVER}}, // 0x18 // API_Runtime ANALOGUE (cudaResViewFormatFloat4 = 0x18)
|
||||
{"CU_RES_VIEW_FORMAT_UNSIGNED_BC1", {"hipResViewFormatUnsignedBlockCompressed1", CONV_TEX, API_DRIVER}}, // 0x19 // API_Runtime ANALOGUE (cudaResViewFormatUnsignedBlockCompressed1 = 0x19)
|
||||
{"CU_RES_VIEW_FORMAT_UNSIGNED_BC2", {"hipResViewFormatUnsignedBlockCompressed2", CONV_TEX, API_DRIVER}}, // 0x1a // API_Runtime ANALOGUE (cudaResViewFormatUnsignedBlockCompressed2 = 0x1a)
|
||||
{"CU_RES_VIEW_FORMAT_UNSIGNED_BC3", {"hipResViewFormatUnsignedBlockCompressed3", CONV_TEX, API_DRIVER}}, // 0x1b // API_Runtime ANALOGUE (cudaResViewFormatUnsignedBlockCompressed3 = 0x1b)
|
||||
{"CU_RES_VIEW_FORMAT_UNSIGNED_BC4", {"hipResViewFormatUnsignedBlockCompressed4", CONV_TEX, API_DRIVER}}, // 0x1c // API_Runtime ANALOGUE (cudaResViewFormatUnsignedBlockCompressed4 = 0x1c)
|
||||
{"CU_RES_VIEW_FORMAT_SIGNED_BC4", {"hipResViewFormatSignedBlockCompressed4", CONV_TEX, API_DRIVER}}, // 0x1d // API_Runtime ANALOGUE (cudaResViewFormatSignedBlockCompressed4 = 0x1d)
|
||||
{"CU_RES_VIEW_FORMAT_UNSIGNED_BC5", {"hipResViewFormatUnsignedBlockCompressed5", CONV_TEX, API_DRIVER}}, // 0x1e // API_Runtime ANALOGUE (cudaResViewFormatUnsignedBlockCompressed5 = 0x1e)
|
||||
{"CU_RES_VIEW_FORMAT_SIGNED_BC5", {"hipResViewFormatSignedBlockCompressed5", CONV_TEX, API_DRIVER}}, // 0x1f // API_Runtime ANALOGUE (cudaResViewFormatSignedBlockCompressed5 = 0x1f)
|
||||
{"CU_RES_VIEW_FORMAT_UNSIGNED_BC6H", {"hipResViewFormatUnsignedBlockCompressed6H", CONV_TEX, API_DRIVER}}, // 0x20 // API_Runtime ANALOGUE (cudaResViewFormatUnsignedBlockCompressed6H = 0x20)
|
||||
{"CU_RES_VIEW_FORMAT_SIGNED_BC6H", {"hipResViewFormatSignedBlockCompressed6H", CONV_TEX, API_DRIVER}}, // 0x21 // API_Runtime ANALOGUE (cudaResViewFormatSignedBlockCompressed6H = 0x21)
|
||||
{"CU_RES_VIEW_FORMAT_UNSIGNED_BC7", {"hipResViewFormatUnsignedBlockCompressed7", CONV_TEX, API_DRIVER}}, // 0x22 // API_Runtime ANALOGUE (cudaResViewFormatUnsignedBlockCompressed7 = 0x22)
|
||||
|
||||
{"CU_SHARED_MEM_CONFIG_DEFAULT_BANK_SIZE", {"hipSharedMemBankSizeDefault", CONV_TYPE, API_DRIVER}},
|
||||
{"CU_SHARED_MEM_CONFIG_FOUR_BYTE_BANK_SIZE", {"hipSharedMemBankSizeFourByte", CONV_TYPE, API_DRIVER}},
|
||||
@@ -981,7 +1040,7 @@ const std::map<llvm::StringRef, hipCounter> CUDA_IDENTIFIER_MAP{
|
||||
{"cuModuleGetFunction", {"hipModuleGetFunction", CONV_MODULE, API_DRIVER}},
|
||||
{"cuModuleGetGlobal_v2", {"hipModuleGetGlobal", CONV_MODULE, API_DRIVER}},
|
||||
{"cuModuleGetSurfRef", {"hipModuleGetSurfRef", CONV_MODULE, API_DRIVER, HIP_UNSUPPORTED}},
|
||||
{"cuModuleGetTexRef", {"hipModuleGetTexRef", CONV_MODULE, API_DRIVER, HIP_UNSUPPORTED}},
|
||||
{"cuModuleGetTexRef", {"hipModuleGetTexRef", CONV_MODULE, API_DRIVER}},
|
||||
{"cuModuleLoad", {"hipModuleLoad", CONV_MODULE, API_DRIVER}},
|
||||
{"cuModuleLoadData", {"hipModuleLoadData", CONV_MODULE, API_DRIVER}},
|
||||
{"cuModuleLoadDataEx", {"hipModuleLoadDataEx", CONV_MODULE, API_DRIVER}},
|
||||
@@ -1047,7 +1106,7 @@ const std::map<llvm::StringRef, hipCounter> CUDA_IDENTIFIER_MAP{
|
||||
{"cuStreamBatchMemOp", {"hipStreamBatchMemOp", CONV_STREAM, API_DRIVER, HIP_UNSUPPORTED}}, // // no API_Runtime ANALOGUE
|
||||
|
||||
// Memory management
|
||||
{"cuArray3DCreate", {"hipArray3DCreate", CONV_MEM, API_DRIVER, HIP_UNSUPPORTED}},
|
||||
{"cuArray3DCreate", {"hipArray3DCreate", CONV_MEM, API_DRIVER}},
|
||||
{"cuArray3DGetDescriptor", {"hipArray3DGetDescriptor", CONV_MEM, API_DRIVER, HIP_UNSUPPORTED}},
|
||||
{"cuArrayCreate", {"hipArrayCreate", CONV_MEM, API_DRIVER, HIP_UNSUPPORTED}},
|
||||
{"cuArrayDestroy", {"hipArrayDestroy", CONV_MEM, API_DRIVER, HIP_UNSUPPORTED}},
|
||||
@@ -1139,12 +1198,12 @@ const std::map<llvm::StringRef, hipCounter> CUDA_IDENTIFIER_MAP{
|
||||
{"cuTexRefGetMipmappedArray", {"hipTexRefGetMipmappedArray", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}},
|
||||
{"cuTexRefSetAddress", {"hipTexRefSetAddress", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}},
|
||||
{"cuTexRefSetAddress2D", {"hipTexRefSetAddress2D", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}},
|
||||
{"cuTexRefSetAddressMode", {"hipTexRefSetAddressMode", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}},
|
||||
{"cuTexRefSetArray", {"hipTexRefSetArray", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}},
|
||||
{"cuTexRefSetAddressMode", {"hipTexRefSetAddressMode", CONV_TEX, API_DRIVER}},
|
||||
{"cuTexRefSetArray", {"hipTexRefSetArray", CONV_TEX, API_DRIVER}},
|
||||
{"cuTexRefSetBorderColor", {"hipTexRefSetBorderColor", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}}, // // no API_Runtime ANALOGUE
|
||||
{"cuTexRefSetFilterMode", {"hipTexRefSetFilterMode", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}},
|
||||
{"cuTexRefSetFlags", {"hipTexRefSetFlags", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}},
|
||||
{"cuTexRefSetFormat", {"hipTexRefSetFormat", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}},
|
||||
{"cuTexRefSetFilterMode", {"hipTexRefSetFilterMode", CONV_TEX, API_DRIVER}},
|
||||
{"cuTexRefSetFlags", {"hipTexRefSetFlags", CONV_TEX, API_DRIVER}},
|
||||
{"cuTexRefSetFormat", {"hipTexRefSetFormat", CONV_TEX, API_DRIVER}},
|
||||
{"cuTexRefSetMaxAnisotropy", {"hipTexRefSetMaxAnisotropy", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}},
|
||||
{"cuTexRefSetMipmapFilterMode", {"hipTexRefSetMipmapFilterMode", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}},
|
||||
{"cuTexRefSetMipmapLevelBias", {"hipTexRefSetMipmapLevelBias", CONV_TEX, API_DRIVER, HIP_UNSUPPORTED}},
|
||||
@@ -1425,7 +1484,7 @@ const std::map<llvm::StringRef, hipCounter> CUDA_IDENTIFIER_MAP{
|
||||
{"cudaMalloc", {"hipMalloc", CONV_MEM, API_RUNTIME}},
|
||||
{"cudaMallocHost", {"hipHostMalloc", CONV_MEM, API_RUNTIME}},
|
||||
{"cudaMallocArray", {"hipMallocArray", CONV_MEM, API_RUNTIME}},
|
||||
{"cudaMalloc3D", {"hipMalloc3D", CONV_MEM, API_RUNTIME, HIP_UNSUPPORTED}},
|
||||
{"cudaMalloc3D", {"hipMalloc3D", CONV_MEM, API_RUNTIME}},
|
||||
{"cudaMalloc3DArray", {"hipMalloc3DArray", CONV_MEM, API_RUNTIME}},
|
||||
{"cudaMallocManaged", {"hipMallocManaged", CONV_MEM, API_RUNTIME, HIP_UNSUPPORTED}},
|
||||
{"cudaMallocMipmappedArray", {"hipMallocMipmappedArray", CONV_MEM, API_RUNTIME, HIP_UNSUPPORTED}},
|
||||
@@ -2004,7 +2063,11 @@ const std::map<llvm::StringRef, hipCounter> CUDA_IDENTIFIER_MAP{
|
||||
{"cublasSetKernelStream", {"hipblasSetKernelStream", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasGetAtomicsMode", {"hipblasGetAtomicsMode", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasSetAtomicsMode", {"hipblasSetAtomicsMode", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
|
||||
{"cublasGetMathMode", {"hipblasGetMathMode", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasSetMathMode", {"hipblasSetMathMode", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
|
||||
|
||||
|
||||
// Blas operations (cublasOperation_t)
|
||||
{"CUBLAS_OP_N", {"HIPBLAS_OP_N", CONV_NUMERIC_LITERAL, API_BLAS}},
|
||||
{"CUBLAS_OP_T", {"HIPBLAS_OP_T", CONV_NUMERIC_LITERAL, API_BLAS}},
|
||||
@@ -2021,20 +2084,20 @@ const std::map<llvm::StringRef, hipCounter> CUDA_IDENTIFIER_MAP{
|
||||
{"CUBLAS_STATUS_NOT_SUPPORTED", {"HIPBLAS_STATUS_INTERNAL_ERROR", CONV_NUMERIC_LITERAL, API_BLAS}},
|
||||
|
||||
// Blas Fill Modes (cublasFillMode_t)
|
||||
{"CUBLAS_FILL_MODE_LOWER", {"HIPBLAS_FILL_MODE_LOWER", CONV_NUMERIC_LITERAL, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"CUBLAS_FILL_MODE_UPPER", {"HIPBLAS_FILL_MODE_UPPER", CONV_NUMERIC_LITERAL, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"CUBLAS_FILL_MODE_LOWER", {"HIPBLAS_FILL_MODE_LOWER", CONV_NUMERIC_LITERAL, API_BLAS}},
|
||||
{"CUBLAS_FILL_MODE_UPPER", {"HIPBLAS_FILL_MODE_UPPER", CONV_NUMERIC_LITERAL, API_BLAS}},
|
||||
|
||||
// Blas Diag Types (cublasDiagType_t)
|
||||
{"CUBLAS_DIAG_NON_UNIT", {"HIPBLAS_DIAG_NON_UNIT", CONV_NUMERIC_LITERAL, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"CUBLAS_DIAG_UNIT", {"HIPBLAS_DIAG_UNIT", CONV_NUMERIC_LITERAL, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"CUBLAS_DIAG_NON_UNIT", {"HIPBLAS_DIAG_NON_UNIT", CONV_NUMERIC_LITERAL, API_BLAS}},
|
||||
{"CUBLAS_DIAG_UNIT", {"HIPBLAS_DIAG_UNIT", CONV_NUMERIC_LITERAL, API_BLAS}},
|
||||
|
||||
// Blas Side Modes (cublasSideMode_t
|
||||
{"CUBLAS_SIDE_LEFT", {"HIPBLAS_SIDE_LEFT", CONV_NUMERIC_LITERAL, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"CUBLAS_SIDE_RIGHT", {"HIPBLAS_SIDE_RIGHT", CONV_NUMERIC_LITERAL, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"CUBLAS_SIDE_LEFT", {"HIPBLAS_SIDE_LEFT", CONV_NUMERIC_LITERAL, API_BLAS}},
|
||||
{"CUBLAS_SIDE_RIGHT", {"HIPBLAS_SIDE_RIGHT", CONV_NUMERIC_LITERAL, API_BLAS}},
|
||||
|
||||
// Blas Pointer Modes (cublasPointerMode_t)
|
||||
{"CUBLAS_POINTER_MODE_HOST", {"HIPBLAS_POINTER_MODE_HOST", CONV_NUMERIC_LITERAL, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"CUBLAS_POINTER_MODE_DEVICE", {"HIPBLAS_POINTER_MODE_DEVICE", CONV_NUMERIC_LITERAL, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"CUBLAS_POINTER_MODE_HOST", {"HIPBLAS_POINTER_MODE_HOST", CONV_NUMERIC_LITERAL, API_BLAS}},
|
||||
{"CUBLAS_POINTER_MODE_DEVICE", {"HIPBLAS_POINTER_MODE_DEVICE", CONV_NUMERIC_LITERAL, API_BLAS}},
|
||||
|
||||
// Blas Atomics Modes (cublasAtomicsMode_t)
|
||||
{"CUBLAS_ATOMICS_NOT_ALLOWED", {"HIPBLAS_ATOMICS_NOT_ALLOWED", CONV_NUMERIC_LITERAL, API_BLAS, HIP_UNSUPPORTED}},
|
||||
@@ -2052,18 +2115,24 @@ const std::map<llvm::StringRef, hipCounter> CUDA_IDENTIFIER_MAP{
|
||||
|
||||
{"cublasSetVector", {"hipblasSetVector", CONV_MATH_FUNC, API_BLAS}},
|
||||
{"cublasGetVector", {"hipblasGetVector", CONV_MATH_FUNC, API_BLAS}},
|
||||
{"cublasSetVectorAsync", {"hipblasSetVectorAsync", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasGetVectorAsync", {"hipblasGetVectorAsync", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
|
||||
{"cublasSetMatrix", {"hipblasSetMatrix", CONV_MATH_FUNC, API_BLAS}},
|
||||
{"cublasGetMatrix", {"hipblasGetMatrix", CONV_MATH_FUNC, API_BLAS}},
|
||||
|
||||
{"cublasGetMatrixAsync", {"hipblasGetMatrixAsync", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasSetMatrixAsync", {"hipblasSetMatrixAsync", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
|
||||
{"cublasXerbla", {"hipblasXerbla", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
|
||||
// NRM2
|
||||
{"cublasSnrm2", {"hipblasSnrm2", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasDnrm2", {"hipblasDnrm2", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasScnrm2", {"hipblasScnrm2", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasDznrm2", {"hipblasDznrm2", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
|
||||
{"cublasNrm2Ex", {"hipblasNrm2Ex", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
|
||||
// DOT
|
||||
{"cublasSdot", {"hipblasSdot", CONV_MATH_FUNC, API_BLAS}},
|
||||
// there is no such a function in CUDA
|
||||
@@ -2112,8 +2181,8 @@ const std::map<llvm::StringRef, hipCounter> CUDA_IDENTIFIER_MAP{
|
||||
{"cublasZswap", {"hipblasZswap", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
|
||||
// AMAX
|
||||
{"cublasIsamax", {"hipblasIsamax", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasIdamax", {"hipblasIdamax", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasIsamax", {"hipblasIsamax", CONV_MATH_FUNC, API_BLAS}},
|
||||
{"cublasIdamax", {"hipblasIdamax", CONV_MATH_FUNC, API_BLAS}},
|
||||
{"cublasIcamax", {"hipblasIcamax", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasIzamax", {"hipblasIzamax", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
|
||||
@@ -2158,7 +2227,7 @@ const std::map<llvm::StringRef, hipCounter> CUDA_IDENTIFIER_MAP{
|
||||
// GEMV
|
||||
{"cublasSgemv", {"hipblasSgemv", CONV_MATH_FUNC, API_BLAS}},
|
||||
// there is no such a function in CUDA
|
||||
{"cublasSgemvBatched", {"hipblasSgemvBatched", CONV_MATH_FUNC, API_BLAS}},
|
||||
{"cublasSgemvBatched", {"hipblasSgemvBatched", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasDgemv", {"hipblasDgemv", CONV_MATH_FUNC, API_BLAS}},
|
||||
{"cublasCgemv", {"hipblasCgemv", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasZgemv", {"hipblasZgemv", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
@@ -2264,14 +2333,25 @@ const std::map<llvm::StringRef, hipCounter> CUDA_IDENTIFIER_MAP{
|
||||
|
||||
{"cublasCgemm", {"hipblasCgemm", CONV_MATH_FUNC, API_BLAS}},
|
||||
{"cublasZgemm", {"hipblasZgemm", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasHgemm", {"hipblasHgemm", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
|
||||
// BATCH GEMM
|
||||
{"cublasSgemmBatched", {"hipblasSgemmBatched", CONV_MATH_FUNC, API_BLAS}},
|
||||
{"cublasDgemmBatched", {"hipblasDgemmBatched", CONV_MATH_FUNC, API_BLAS}},
|
||||
{"cublasHgemmBatched", {"hipblasHgemmBatched", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
|
||||
{"cublasCgemmBatched", {"hipblasCgemmBatched", CONV_MATH_FUNC, API_BLAS}},
|
||||
{"cublasSgemmStridedBatched", {"hipblasSgemmStridedBatched", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasDgemmStridedBatched", {"hipblasDgemmStridedBatched", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
|
||||
{"cublasCgemmBatched", {"hipblasCgemmBatched", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasCgemm3mBatched", {"hipblasCgemm3mBatched", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasZgemmBatched", {"hipblasZgemmBatched", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
|
||||
{"cublasCgemmStridedBatched", {"hipblasCgemmStridedBatched", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasCgemm3mStridedBatched", {"hipblasCgemm3mStridedBatched", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasZgemmStridedBatched", {"hipblasZgemmStridedBatched", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasHgemmStridedBatched", {"hipblasHgemmStridedBatched", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
|
||||
// SYRK
|
||||
{"cublasSsyrk", {"hipblasSsyrk", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasDsyrk", {"hipblasDsyrk", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
@@ -2294,7 +2374,6 @@ const std::map<llvm::StringRef, hipCounter> CUDA_IDENTIFIER_MAP{
|
||||
{"cublasCsyrkx", {"hipblasCsyrkx", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasZsyrkx", {"hipblasZsyrkx", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
|
||||
|
||||
// HER2K
|
||||
{"cublasCher2k", {"hipblasCher2k", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasZher2k", {"hipblasZher2k", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
@@ -2333,8 +2412,8 @@ const std::map<llvm::StringRef, hipCounter> CUDA_IDENTIFIER_MAP{
|
||||
|
||||
// ------------------------ CUBLAS BLAS - like extension (cublas_api.h)
|
||||
// GEAM
|
||||
{"cublasSgeam", {"hipblasSgeam", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasDgeam", {"hipblasDgeam", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasSgeam", {"hipblasSgeam", CONV_MATH_FUNC, API_BLAS}},
|
||||
{"cublasDgeam", {"hipblasDgeam", CONV_MATH_FUNC, API_BLAS}},
|
||||
{"cublasCgeam", {"hipblasCgeam", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasZgeam", {"hipblasZgeam", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
|
||||
@@ -2405,8 +2484,8 @@ const std::map<llvm::StringRef, hipCounter> CUDA_IDENTIFIER_MAP{
|
||||
{"cublasGetVersion_v2", {"hipblasGetVersion", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasSetStream_v2", {"hipblasSetStream", CONV_MATH_FUNC, API_BLAS}},
|
||||
{"cublasGetStream_v2", {"hipblasGetStream", CONV_MATH_FUNC, API_BLAS}},
|
||||
{"cublasGetPointerMode_v2", {"hipblasGetPointerMode", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasSetPointerMode_v2", {"hipblasSetPointerMode", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasGetPointerMode_v2", {"hipblasGetPointerMode", CONV_MATH_FUNC, API_BLAS}},
|
||||
{"cublasSetPointerMode_v2", {"hipblasSetPointerMode", CONV_MATH_FUNC, API_BLAS}},
|
||||
|
||||
// GEMV
|
||||
{"cublasSgemv_v2", {"hipblasSgemv", CONV_MATH_FUNC, API_BLAS}},
|
||||
@@ -2478,7 +2557,7 @@ const std::map<llvm::StringRef, hipCounter> CUDA_IDENTIFIER_MAP{
|
||||
|
||||
// GER
|
||||
{"cublasSger_v2", {"hipblasSger", CONV_MATH_FUNC, API_BLAS}},
|
||||
{"cublasDger_v2", {"hipblasDger", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasDger_v2", {"hipblasDger", CONV_MATH_FUNC, API_BLAS}},
|
||||
{"cublasCgeru_v2", {"hipblasCgeru", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasCgerc_v2", {"hipblasCgerc", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasZgeru_v2", {"hipblasZgeru", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
@@ -2515,13 +2594,22 @@ const std::map<llvm::StringRef, hipCounter> CUDA_IDENTIFIER_MAP{
|
||||
// Blas3 (v2) Routines
|
||||
// GEMM
|
||||
{"cublasSgemm_v2", {"hipblasSgemm", CONV_MATH_FUNC, API_BLAS}},
|
||||
{"cublasDgemm_v2", {"hipblasDgemm", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasDgemm_v2", {"hipblasDgemm", CONV_MATH_FUNC, API_BLAS}},
|
||||
|
||||
{"cublasCgemm_v2", {"hipblasCgemm", CONV_MATH_FUNC, API_BLAS}},
|
||||
{"cublasCgemm3m", {"hipblasCgemm3m", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasCgemm3mEx", {"hipblasCgemm3mEx", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
|
||||
{"cublasZgemm_v2", {"hipblasZgemm", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasZgemm3m", {"hipblasZgemm3m", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
|
||||
//IO in FP16 / FP32, computation in float
|
||||
{"cublasSgemmEx", {"hipblasSgemmEx", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasGemmEx", {"hipblasGemmEx", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
// IO in Int8 complex/cuComplex, computation in cuComplex
|
||||
{"cublasCgemmEx", {"hipblasCgemmEx", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
|
||||
{"cublasUint8gemmBias", {"hipblasUint8gemmBias", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
|
||||
// SYRK
|
||||
{"cublasSsyrk_v2", {"hipblasSsyrk", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
@@ -2529,8 +2617,17 @@ const std::map<llvm::StringRef, hipCounter> CUDA_IDENTIFIER_MAP{
|
||||
{"cublasCsyrk_v2", {"hipblasCsyrk", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasZsyrk_v2", {"hipblasZsyrk", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
|
||||
// IO in Int8 complex/cuComplex, computation in cuComplex
|
||||
{"cublasCsyrkEx", {"hipblasCsyrkEx", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
// IO in Int8 complex/cuComplex, computation in cuComplex, Gaussian math
|
||||
{"cublasCsyrk3mEx", {"hipblasCsyrk3mEx", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
|
||||
// HERK
|
||||
{"cublasCherk_v2", {"hipblasCherk", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
// IO in Int8 complex/cuComplex, computation in cuComplex
|
||||
{"cublasCherkEx", {"hipblasCherkEx", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
// IO in Int8 complex/cuComplex, computation in cuComplex, Gaussian math
|
||||
{"cublasCherk3mEx", {"hipblasCherk3mEx", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasZherk_v2", {"hipblasZherk", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
|
||||
// SYR2K
|
||||
@@ -2572,6 +2669,9 @@ const std::map<llvm::StringRef, hipCounter> CUDA_IDENTIFIER_MAP{
|
||||
{"cublasDznrm2_v2", {"hipblasDznrm2", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
|
||||
// DOT
|
||||
{"cublasDotEx", {"hipblasDotEx", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasDotcEx", {"hipblasDotcEx", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
|
||||
{"cublasSdot_v2", {"hipblasSdot", CONV_MATH_FUNC, API_BLAS}},
|
||||
{"cublasDdot_v2", {"hipblasDdot", CONV_MATH_FUNC, API_BLAS}},
|
||||
|
||||
@@ -2581,6 +2681,7 @@ const std::map<llvm::StringRef, hipCounter> CUDA_IDENTIFIER_MAP{
|
||||
{"cublasZdotc_v2", {"hipblasZdotc", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
|
||||
// SCAL
|
||||
{"cublasScalEx", {"hipblasScalEx", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasSscal_v2", {"hipblasSscal", CONV_MATH_FUNC, API_BLAS}},
|
||||
{"cublasDscal_v2", {"hipblasDscal", CONV_MATH_FUNC, API_BLAS}},
|
||||
{"cublasCscal_v2", {"hipblasCscal", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
@@ -2589,6 +2690,7 @@ const std::map<llvm::StringRef, hipCounter> CUDA_IDENTIFIER_MAP{
|
||||
{"cublasZdscal_v2", {"hipblasZdscal", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
|
||||
// AXPY
|
||||
{"cublasAxpyEx", {"hipblasAxpyEx", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasSaxpy_v2", {"hipblasSaxpy", CONV_MATH_FUNC, API_BLAS}},
|
||||
{"cublasDaxpy_v2", {"hipblasDaxpy", CONV_MATH_FUNC, API_BLAS}},
|
||||
{"cublasCaxpy_v2", {"hipblasCaxpy", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
@@ -2607,8 +2709,8 @@ const std::map<llvm::StringRef, hipCounter> CUDA_IDENTIFIER_MAP{
|
||||
{"cublasZswap_v2", {"hipblasZswap", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
|
||||
// AMAX
|
||||
{"cublasIsamax_v2", {"hipblasIsamax", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasIdamax_v2", {"hipblasIdamax", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasIsamax_v2", {"hipblasIsamax", CONV_MATH_FUNC, API_BLAS}},
|
||||
{"cublasIdamax_v2", {"hipblasIdamax", CONV_MATH_FUNC, API_BLAS}},
|
||||
{"cublasIcamax_v2", {"hipblasIcamax", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasIzamax_v2", {"hipblasIzamax", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
|
||||
@@ -2644,7 +2746,121 @@ const std::map<llvm::StringRef, hipCounter> CUDA_IDENTIFIER_MAP{
|
||||
|
||||
// ROTMG
|
||||
{"cublasSrotmg_v2", {"hipblasSrotmg", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
{"cublasDrotmg_v2", {"hipblasDrotmg", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}}
|
||||
{"cublasDrotmg_v2", {"hipblasDrotmg", CONV_MATH_FUNC, API_BLAS, HIP_UNSUPPORTED}},
|
||||
|
||||
///////////////////////////// cuRAND /////////////////////////////
|
||||
// RAND function call status types (enum curandStatus)
|
||||
{"CURAND_STATUS_SUCCESS", {"HIPRAND_STATUS_SUCCESS", CONV_NUMERIC_LITERAL, API_RAND}},
|
||||
{"CURAND_STATUS_VERSION_MISMATCH", {"HIPRAND_STATUS_VERSION_MISMATCH", CONV_NUMERIC_LITERAL, API_RAND}},
|
||||
{"CURAND_STATUS_NOT_INITIALIZED", {"HIPRAND_STATUS_NOT_INITIALIZED", CONV_NUMERIC_LITERAL, API_RAND}},
|
||||
{"CURAND_STATUS_ALLOCATION_FAILED", {"HIPRAND_STATUS_ALLOCATION_FAILED", CONV_NUMERIC_LITERAL, API_RAND}},
|
||||
{"CURAND_STATUS_TYPE_ERROR", {"HIPRAND_STATUS_TYPE_ERROR", CONV_NUMERIC_LITERAL, API_RAND}},
|
||||
{"CURAND_STATUS_OUT_OF_RANGE", {"HIPRAND_STATUS_OUT_OF_RANGE", CONV_NUMERIC_LITERAL, API_RAND}},
|
||||
{"CURAND_STATUS_LENGTH_NOT_MULTIPLE", {"HIPRAND_STATUS_LENGTH_NOT_MULTIPLE", CONV_NUMERIC_LITERAL, API_RAND}},
|
||||
{"CURAND_STATUS_DOUBLE_PRECISION_REQUIRED", {"HIPRAND_STATUS_DOUBLE_PRECISION_REQUIRED", CONV_NUMERIC_LITERAL, API_RAND}},
|
||||
{"CURAND_STATUS_LAUNCH_FAILURE", {"HIPRAND_STATUS_LAUNCH_FAILURE", CONV_NUMERIC_LITERAL, API_RAND}},
|
||||
{"CURAND_STATUS_PREEXISTING_FAILURE", {"HIPRAND_STATUS_PREEXISTING_FAILURE", CONV_NUMERIC_LITERAL, API_RAND}},
|
||||
{"CURAND_STATUS_INITIALIZATION_FAILED", {"HIPRAND_STATUS_INITIALIZATION_FAILED", CONV_NUMERIC_LITERAL, API_RAND}},
|
||||
{"CURAND_STATUS_ARCH_MISMATCH", {"HIPRAND_STATUS_ARCH_MISMATCH", CONV_NUMERIC_LITERAL, API_RAND}},
|
||||
{"CURAND_STATUS_INTERNAL_ERROR", {"HIPRAND_STATUS_INTERNAL_ERROR", CONV_NUMERIC_LITERAL, API_RAND}},
|
||||
|
||||
// RAND generator types (enum curandRngType)
|
||||
{"CURAND_RNG_TEST", {"HIPRAND_RNG_TEST", CONV_NUMERIC_LITERAL, API_RAND}},
|
||||
{"CURAND_RNG_PSEUDO_DEFAULT", {"HIPRAND_RNG_PSEUDO_DEFAULT", CONV_NUMERIC_LITERAL, API_RAND}},
|
||||
{"CURAND_RNG_PSEUDO_XORWOW", {"HIPRAND_RNG_PSEUDO_XORWOW", CONV_NUMERIC_LITERAL, API_RAND}},
|
||||
{"CURAND_RNG_PSEUDO_MRG32K3A", {"HIPRAND_RNG_PSEUDO_MRG32K3A", CONV_NUMERIC_LITERAL, API_RAND}},
|
||||
{"CURAND_RNG_PSEUDO_MTGP32", {"HIPRAND_RNG_PSEUDO_MTGP32", CONV_NUMERIC_LITERAL, API_RAND}},
|
||||
{"CURAND_RNG_PSEUDO_MT19937", {"HIPRAND_RNG_PSEUDO_MT19937", CONV_NUMERIC_LITERAL, API_RAND}},
|
||||
{"CURAND_RNG_PSEUDO_PHILOX4_32_10", {"HIPRAND_RNG_PSEUDO_PHILOX4_32_10", CONV_NUMERIC_LITERAL, API_RAND}},
|
||||
{"CURAND_RNG_QUASI_DEFAULT", {"HIPRAND_RNG_QUASI_DEFAULT", CONV_NUMERIC_LITERAL, API_RAND}},
|
||||
{"CURAND_RNG_QUASI_SOBOL32", {"HIPRAND_RNG_QUASI_SOBOL32", CONV_NUMERIC_LITERAL, API_RAND}},
|
||||
{"CURAND_RNG_QUASI_SCRAMBLED_SOBOL32", {"HIPRAND_RNG_QUASI_SCRAMBLED_SOBOL32", CONV_NUMERIC_LITERAL, API_RAND}},
|
||||
{"CURAND_RNG_QUASI_SOBOL64", {"HIPRAND_RNG_QUASI_SOBOL64", CONV_NUMERIC_LITERAL, API_RAND}},
|
||||
{"CURAND_RNG_QUASI_SCRAMBLED_SOBOL64", {"HIPRAND_RNG_QUASI_SCRAMBLED_SOBOL64", CONV_NUMERIC_LITERAL, API_RAND}},
|
||||
|
||||
// RAND ordering of results in memory (enum curandOrdering)
|
||||
{"CURAND_ORDERING_PSEUDO_BEST", {"HIPRAND_ORDERING_PSEUDO_BEST", CONV_NUMERIC_LITERAL, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"CURAND_ORDERING_PSEUDO_DEFAULT", {"HIPRAND_ORDERING_PSEUDO_DEFAULT", CONV_NUMERIC_LITERAL, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"CURAND_ORDERING_PSEUDO_SEEDED", {"HIPRAND_ORDERING_PSEUDO_SEEDED", CONV_NUMERIC_LITERAL, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"CURAND_ORDERING_QUASI_DEFAULT", {"HIPRAND_ORDERING_QUASI_DEFAULT", CONV_NUMERIC_LITERAL, API_RAND, HIP_UNSUPPORTED}},
|
||||
|
||||
// RAND choice of direction vector set (enum curandDirectionVectorSet)
|
||||
{"CURAND_DIRECTION_VECTORS_32_JOEKUO6", {"HIPRAND_DIRECTION_VECTORS_32_JOEKUO6", CONV_NUMERIC_LITERAL, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"CURAND_SCRAMBLED_DIRECTION_VECTORS_32_JOEKUO6", {"HIPRAND_SCRAMBLED_DIRECTION_VECTORS_32_JOEKUO6", CONV_NUMERIC_LITERAL, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"CURAND_DIRECTION_VECTORS_64_JOEKUO6", {"HIPRAND_DIRECTION_VECTORS_64_JOEKUO6", CONV_NUMERIC_LITERAL, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"CURAND_SCRAMBLED_DIRECTION_VECTORS_64_JOEKUO6", {"HIPRAND_SCRAMBLED_DIRECTION_VECTORS_64_JOEKUO6", CONV_NUMERIC_LITERAL, API_RAND, HIP_UNSUPPORTED}},
|
||||
|
||||
// RAND method (enum curandMethod)
|
||||
{"CURAND_CHOOSE_BEST", {"HIPRAND_CHOOSE_BEST", CONV_NUMERIC_LITERAL, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"CURAND_ITR", {"HIPRAND_ITR", CONV_NUMERIC_LITERAL, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"CURAND_KNUTH", {"HIPRAND_KNUTH", CONV_NUMERIC_LITERAL, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"CURAND_HITR", {"HIPRAND_HITR", CONV_NUMERIC_LITERAL, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"CURAND_M1", {"HIPRAND_M1", CONV_NUMERIC_LITERAL, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"CURAND_M2", {"HIPRAND_M2", CONV_NUMERIC_LITERAL, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"CURAND_BINARY_SEARCH", {"HIPRAND_BINARY_SEARCH", CONV_NUMERIC_LITERAL, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"CURAND_DISCRETE_GAUSS", {"HIPRAND_DISCRETE_GAUSS", CONV_NUMERIC_LITERAL, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"CURAND_REJECTION", {"HIPRAND_REJECTION", CONV_NUMERIC_LITERAL, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"CURAND_DEVICE_API", {"HIPRAND_DEVICE_API", CONV_NUMERIC_LITERAL, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"CURAND_FAST_REJECTION", {"HIPRAND_FAST_REJECTION", CONV_NUMERIC_LITERAL, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"CURAND_3RD", {"HIPRAND_3RD", CONV_NUMERIC_LITERAL, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"CURAND_DEFINITION", {"HIPRAND_DEFINITION", CONV_NUMERIC_LITERAL, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"CURAND_POISSON", {"HIPRAND_POISSON", CONV_NUMERIC_LITERAL, API_RAND, HIP_UNSUPPORTED}},
|
||||
|
||||
// RAND Host functions
|
||||
{"curandCreateGenerator", {"hiprandCreateGenerator", CONV_MATH_FUNC, API_RAND}},
|
||||
{"curandCreateGeneratorHost", {"hiprandCreateGeneratorHost", CONV_MATH_FUNC, API_RAND}},
|
||||
{"curandCreatePoissonDistribution", {"hiprandCreatePoissonDistribution", CONV_MATH_FUNC, API_RAND}},
|
||||
{"curandDestroyDistribution", {"hiprandDestroyDistribution", CONV_MATH_FUNC, API_RAND}},
|
||||
{"curandDestroyGenerator", {"hiprandDestroyGenerator", CONV_MATH_FUNC, API_RAND}},
|
||||
{"curandGenerate", {"hiprandGenerate", CONV_MATH_FUNC, API_RAND}},
|
||||
{"curandGenerateLogNormal", {"hiprandGenerateLogNormal", CONV_MATH_FUNC, API_RAND}},
|
||||
{"curandGenerateLogNormalDouble", {"hiprandGenerateLogNormalDouble", CONV_MATH_FUNC, API_RAND}},
|
||||
{"curandGenerateLongLong", {"hiprandGenerateLongLong", CONV_MATH_FUNC, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"curandGenerateNormal", {"hiprandGenerateNormal", CONV_MATH_FUNC, API_RAND}},
|
||||
{"curandGenerateNormalDouble", {"hiprandGenerateNormalDouble", CONV_MATH_FUNC, API_RAND}},
|
||||
{"curandGeneratePoisson", {"hiprandGeneratePoisson", CONV_MATH_FUNC, API_RAND}},
|
||||
{"curandGenerateSeeds", {"hiprandGenerateSeeds", CONV_MATH_FUNC, API_RAND}},
|
||||
{"curandGenerateUniform", {"hiprandGenerateUniform", CONV_MATH_FUNC, API_RAND}},
|
||||
{"curandGenerateUniformDouble", {"hiprandGenerateUniformDouble", CONV_MATH_FUNC, API_RAND}},
|
||||
{"curandGetDirectionVectors32", {"hiprandGetDirectionVectors32", CONV_MATH_FUNC, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"curandGetDirectionVectors64", {"hiprandGetDirectionVectors64", CONV_MATH_FUNC, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"curandGetProperty", {"hiprandGetProperty", CONV_MATH_FUNC, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"curandGetScrambleConstants32", {"hiprandGetScrambleConstants32", CONV_MATH_FUNC, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"curandGetScrambleConstants64", {"hiprandGetScrambleConstants64", CONV_MATH_FUNC, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"curandGetVersion", {"hiprandGetVersion", CONV_MATH_FUNC, API_RAND}},
|
||||
{"curandSetGeneratorOffset", {"hiprandSetGeneratorOffset", CONV_MATH_FUNC, API_RAND}},
|
||||
{"curandSetGeneratorOrdering", {"hiprandSetGeneratorOrdering", CONV_MATH_FUNC, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"curandSetPseudoRandomGeneratorSeed", {"hiprandSetPseudoRandomGeneratorSeed", CONV_MATH_FUNC, API_RAND}},
|
||||
{"curandSetQuasiRandomGeneratorDimensions", {"hiprandSetQuasiRandomGeneratorDimensions", CONV_MATH_FUNC, API_RAND}},
|
||||
{"curandSetStream", {"hiprandSetStream", CONV_MATH_FUNC, API_RAND}},
|
||||
|
||||
// RAND Device functions
|
||||
{"curand", {"hiprand", CONV_DEVICE_FUNC, API_RAND}},
|
||||
{"curand_init", {"hiprand_init", CONV_DEVICE_FUNC, API_RAND}},
|
||||
{"curand_log_normal", {"hiprand_log_normal", CONV_DEVICE_FUNC, API_RAND}},
|
||||
{"curand_log_normal_double", {"hiprand_log_normal_double", CONV_DEVICE_FUNC, API_RAND}},
|
||||
{"curand_log_normal2", {"hiprand_log_normal2", CONV_DEVICE_FUNC, API_RAND}},
|
||||
{"curand_log_normal2_double", {"hiprand_log_normal2_double", CONV_DEVICE_FUNC, API_RAND}},
|
||||
{"curand_log_normal4", {"hiprand_log_normal4", CONV_DEVICE_FUNC, API_RAND}},
|
||||
{"curand_log_normal4_double", {"hiprand_log_normal4_double", CONV_DEVICE_FUNC, API_RAND}},
|
||||
{"curand_mtgp32_single", {"hiprand_mtgp32_single", CONV_DEVICE_FUNC, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"curand_mtgp32_single_specific", {"hiprand_mtgp32_single_specific", CONV_DEVICE_FUNC, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"curand_mtgp32_specific", {"hiprand_mtgp32_specific", CONV_DEVICE_FUNC, API_RAND, HIP_UNSUPPORTED}},
|
||||
{"curand_normal", {"hiprand_normal", CONV_DEVICE_FUNC, API_RAND}},
|
||||
{"curand_normal_double", {"hiprand_normal_double", CONV_DEVICE_FUNC, API_RAND}},
|
||||
{"curand_normal2", {"hiprand_normal2", CONV_DEVICE_FUNC, API_RAND}},
|
||||
{"curand_normal2_double", {"hiprand_normal2_double", CONV_DEVICE_FUNC, API_RAND}},
|
||||
{"curand_normal4", {"hiprand_normal4", CONV_DEVICE_FUNC, API_RAND}},
|
||||
{"curand_normal4_double", {"hiprand_normal4_double", CONV_DEVICE_FUNC, API_RAND}},
|
||||
{"curand_uniform2_double", {"hiprand_uniform2_double", CONV_DEVICE_FUNC, API_RAND}},
|
||||
{"curand_uniform4", {"hiprand_uniform4", CONV_DEVICE_FUNC, API_RAND}},
|
||||
{"curand_uniform4_double", {"hiprand_uniform4_double", CONV_DEVICE_FUNC, API_RAND}},
|
||||
{"curand_discrete", {"hiprand_discrete4", CONV_DEVICE_FUNC, API_RAND}},
|
||||
{"curand_discrete4", {"hiprand_discrete4", CONV_DEVICE_FUNC, API_RAND}},
|
||||
{"curand_poisson", {"hiprand_poisson", CONV_DEVICE_FUNC, API_RAND}},
|
||||
{"curand_poisson4", {"hiprand_poisson4", CONV_DEVICE_FUNC, API_RAND}},
|
||||
{"curand_Philox4x32_10", {"hiprand_Philox4x32_10", CONV_DEVICE_FUNC, API_RAND, HIP_UNSUPPORTED}},
|
||||
// unchanged function names: skipahead, skipahead_sequence, skipahead_subsequence
|
||||
};
|
||||
|
||||
const std::map<llvm::StringRef, hipCounter>& CUDA_RENAMES_MAP() {
|
||||
|
||||
@@ -149,20 +149,38 @@ void HipifyAction::InclusionDirective(clang::SourceLocation hash_loc,
|
||||
if (!SM.isWrittenInMainFile(hash_loc)) {
|
||||
return;
|
||||
}
|
||||
if (!firstHeader) {
|
||||
firstHeader = true;
|
||||
firstHeaderLoc = hash_loc;
|
||||
}
|
||||
|
||||
const auto found = CUDA_INCLUDE_MAP.find(file_name);
|
||||
if (found == CUDA_INCLUDE_MAP.end()) {
|
||||
// Not a CUDA include - don't touch it.
|
||||
return;
|
||||
}
|
||||
|
||||
// Special-casing to avoid duplication of the hip_runtime include.
|
||||
bool secondMainInclude = false;
|
||||
if (found->second.hipName == "hip/hip_runtime.h") {
|
||||
if (insertedRuntimeHeader) {
|
||||
secondMainInclude = true;
|
||||
if (found->second.countType == CONV_INCLUDE_CUDA_MAIN_H) {
|
||||
switch (found->second.countApiType) {
|
||||
case API_DRIVER:
|
||||
case API_RUNTIME:
|
||||
if (insertedRuntimeHeader) {
|
||||
secondMainInclude = true;
|
||||
break;
|
||||
}
|
||||
insertedRuntimeHeader = true;
|
||||
break;
|
||||
case API_BLAS:
|
||||
if (insertedBLASHeader) {
|
||||
secondMainInclude = true;
|
||||
break;
|
||||
}
|
||||
insertedBLASHeader = true;
|
||||
break;
|
||||
default:
|
||||
break;
|
||||
}
|
||||
insertedRuntimeHeader = true;
|
||||
}
|
||||
|
||||
Statistics::current().incrementCounter(found->second, file_name.str());
|
||||
@@ -178,15 +196,15 @@ void HipifyAction::InclusionDirective(clang::SourceLocation hash_loc,
|
||||
|
||||
// Keep the same include type that the user gave.
|
||||
if (!secondMainInclude) {
|
||||
clang::SmallString<128> includeBuffer;
|
||||
if (is_angled) {
|
||||
newInclude = llvm::Twine("<" + found->second.hipName + ">").toStringRef(includeBuffer);
|
||||
} else {
|
||||
newInclude = llvm::Twine("\"" + found->second.hipName + "\"").toStringRef(includeBuffer);
|
||||
}
|
||||
clang::SmallString<128> includeBuffer;
|
||||
if (is_angled) {
|
||||
newInclude = llvm::Twine("<" + found->second.hipName + ">").toStringRef(includeBuffer);
|
||||
} else {
|
||||
newInclude = llvm::Twine("\"" + found->second.hipName + "\"").toStringRef(includeBuffer);
|
||||
}
|
||||
} else {
|
||||
// hashLoc is location of the '#', thus replacing the whole include directive by empty newInclude starting with '#'.
|
||||
sl = hash_loc;
|
||||
// hashLoc is location of the '#', thus replacing the whole include directive by empty newInclude starting with '#'.
|
||||
sl = hash_loc;
|
||||
}
|
||||
const char *B = SM.getCharacterData(sl);
|
||||
const char *E = SM.getCharacterData(filename_range.getEnd());
|
||||
@@ -194,6 +212,22 @@ void HipifyAction::InclusionDirective(clang::SourceLocation hash_loc,
|
||||
insertReplacement(Rep, clang::FullSourceLoc{sl, SM});
|
||||
}
|
||||
|
||||
void HipifyAction::PragmaDirective(clang::SourceLocation Loc, clang::PragmaIntroducerKind Introducer) {
|
||||
if (pragmaOnce) {
|
||||
return;
|
||||
}
|
||||
clang::SourceManager& SM = getCompilerInstance().getSourceManager();
|
||||
if (!SM.isWrittenInMainFile(Loc)) {
|
||||
return;
|
||||
}
|
||||
clang::Preprocessor& PP = getCompilerInstance().getPreprocessor();
|
||||
const clang::Token tok = PP.LookAhead(0);
|
||||
StringRef Text(SM.getCharacterData(tok.getLocation()), tok.getLength());
|
||||
if (Text == "once") {
|
||||
pragmaOnce = true;
|
||||
pragmaOnceLoc = PP.LookAhead(1).getLocation();
|
||||
}
|
||||
}
|
||||
|
||||
bool HipifyAction::cudaLaunchKernel(const clang::ast_matchers::MatchFinder::MatchResult& Result) {
|
||||
StringRef refName = "cudaLaunchKernel";
|
||||
@@ -339,10 +373,16 @@ void HipifyAction::EndSourceFileAction() {
|
||||
// implicitly included by the compiler. Instead, we _delete_ CUDA headers, and unconditionally insert
|
||||
// one copy of the hip include into every file.
|
||||
clang::SourceManager& SM = getCompilerInstance().getSourceManager();
|
||||
|
||||
clang::SourceLocation sl = SM.getLocForStartOfFile(SM.getMainFileID());
|
||||
clang::SourceLocation sl;
|
||||
if (pragmaOnce) {
|
||||
sl = pragmaOnceLoc;
|
||||
} else if (firstHeader) {
|
||||
sl = firstHeaderLoc;
|
||||
} else {
|
||||
sl = SM.getLocForStartOfFile(SM.getMainFileID());
|
||||
}
|
||||
clang::FullSourceLoc fullSL(sl, SM);
|
||||
ct::Replacement Rep(SM, sl, 0, "#include <hip/hip_runtime.h>\n");
|
||||
ct::Replacement Rep(SM, sl, 0, "\n#include <hip/hip_runtime.h>\n");
|
||||
insertReplacement(Rep, fullSL);
|
||||
}
|
||||
|
||||
@@ -367,6 +407,10 @@ public:
|
||||
const clang::Module* imported) override {
|
||||
hipifyAction.InclusionDirective(hash_loc, include_token, file_name, is_angled, filename_range, file, search_path, relative_path, imported);
|
||||
}
|
||||
|
||||
void PragmaDirective(clang::SourceLocation Loc, clang::PragmaIntroducerKind Introducer) override {
|
||||
hipifyAction.PragmaDirective(Loc, Introducer);
|
||||
}
|
||||
};
|
||||
|
||||
}
|
||||
|
||||
@@ -23,6 +23,11 @@ private:
|
||||
// not, we insert it at the top of the file when we finish processing it.
|
||||
// This approach means we do the best it's possible to do w.r.t preserving the user's include order.
|
||||
bool insertedRuntimeHeader = false;
|
||||
bool insertedBLASHeader = false;
|
||||
bool firstHeader = false;
|
||||
bool pragmaOnce = false;
|
||||
clang::SourceLocation firstHeaderLoc;
|
||||
clang::SourceLocation pragmaOnceLoc;
|
||||
|
||||
/**
|
||||
* Rewrite a string literal to refer to hip, not CUDA.
|
||||
@@ -57,6 +62,11 @@ public:
|
||||
StringRef relative_path,
|
||||
const clang::Module *imported);
|
||||
|
||||
/**
|
||||
* Called by the preprocessor for each pragma directive during the non-raw lexing pass.
|
||||
*/
|
||||
void PragmaDirective(clang::SourceLocation Loc, clang::PragmaIntroducerKind Introducer);
|
||||
|
||||
protected:
|
||||
/**
|
||||
* Add a Replacement for the current file. These will all be applied after executing the FrontendAction.
|
||||
|
||||
@@ -5,7 +5,7 @@
|
||||
|
||||
|
||||
const char *counterNames[NUM_CONV_TYPES] = {
|
||||
"version", "init", "device", "mem", "kern", "coord_func", "math_func",
|
||||
"version", "init", "device", "mem", "kern", "coord_func", "math_func", "device_func",
|
||||
"special_func", "stream", "event", "occupancy", "ctx", "peer", "module",
|
||||
"cache", "exec", "err", "def", "tex", "gl", "graphics",
|
||||
"surface", "jit", "d3d9", "d3d10", "d3d11", "vdpau", "egl",
|
||||
@@ -14,7 +14,7 @@ const char *counterNames[NUM_CONV_TYPES] = {
|
||||
};
|
||||
|
||||
const char *apiNames[NUM_API_TYPES] = {
|
||||
"CUDA Driver API", "CUDA RT API", "CUBLAS API"
|
||||
"CUDA Driver API", "CUDA RT API", "CUBLAS API", "CURAND API"
|
||||
};
|
||||
|
||||
namespace {
|
||||
|
||||
@@ -18,6 +18,7 @@ enum ConvTypes {
|
||||
CONV_KERN,
|
||||
CONV_COORD_FUNC,
|
||||
CONV_MATH_FUNC,
|
||||
CONV_DEVICE_FUNC,
|
||||
CONV_SPECIAL_FUNC,
|
||||
CONV_STREAM,
|
||||
CONV_EVENT,
|
||||
@@ -54,6 +55,7 @@ enum ApiTypes {
|
||||
API_DRIVER = 0,
|
||||
API_RUNTIME,
|
||||
API_BLAS,
|
||||
API_RAND,
|
||||
API_LAST
|
||||
};
|
||||
constexpr int NUM_API_TYPES = (int) ApiTypes::API_LAST;
|
||||
|
||||
@@ -88,7 +88,7 @@ namespace hip_impl
|
||||
std::copy_n(it, sizeof(y.cbuf), y.cbuf);
|
||||
it += sizeof(y.cbuf);
|
||||
|
||||
y.triple.insert(y.triple.cend(), it, it + y.triple_sz);
|
||||
y.triple.assign(it, it + y.triple_sz);
|
||||
|
||||
std::copy_n(
|
||||
f + y.offset, y.bundle_sz, std::back_inserter(y.blob));
|
||||
@@ -150,9 +150,9 @@ namespace hip_impl
|
||||
};
|
||||
|
||||
// CREATORS
|
||||
template<typename I>
|
||||
Bundled_code_header::Bundled_code_header(I f, I l) : Bundled_code_header{}
|
||||
template<typename RandomAccessIterator>
|
||||
Bundled_code_header::Bundled_code_header(RandomAccessIterator f, RandomAccessIterator l) : Bundled_code_header{}
|
||||
{
|
||||
read(f, l, *this);
|
||||
}
|
||||
} // Namespace hip_impl.
|
||||
} // Namespace hip_impl.
|
||||
|
||||
@@ -8,12 +8,12 @@ if(DOXYGEN_EXE)
|
||||
install(DIRECTORY RuntimeAPI/html DESTINATION docs/docs/RuntimeAPI)
|
||||
endif()
|
||||
|
||||
find_program(GRIP_EXE grip)
|
||||
if(GRIP_EXE)
|
||||
add_custom_target(convert_md_to_html ALL
|
||||
COMMAND @hip_SOURCE_DIR@/packaging/convert_md_to_html.sh @hip_SOURCE_DIR@ ${PROJECT_BINARY_DIR}/md2html)
|
||||
install(DIRECTORY md2html/ DESTINATION docs)
|
||||
endif()
|
||||
#find_program(GRIP_EXE grip)
|
||||
#if(GRIP_EXE)
|
||||
# add_custom_target(convert_md_to_html ALL
|
||||
# COMMAND @hip_SOURCE_DIR@/packaging/convert_md_to_html.sh @hip_SOURCE_DIR@ ${PROJECT_BINARY_DIR}/md2html)
|
||||
# install(DIRECTORY md2html/ DESTINATION docs)
|
||||
#endif()
|
||||
|
||||
#############################
|
||||
# Packaging steps
|
||||
|
||||
@@ -16,7 +16,6 @@
|
||||
|
||||
template<typename T>
|
||||
__global__ void axpy(T a, T *x, T *y) {
|
||||
// CHECK: y[hipThreadIdx_x] = a * x[hipThreadIdx_x];
|
||||
y[threadIdx.x] = a * x[threadIdx.x];
|
||||
}
|
||||
|
||||
|
||||
@@ -1,242 +0,0 @@
|
||||
// RUN: %run_test hipify "%s" "%t" %cuda_args
|
||||
/*
|
||||
* Copyright 1993-2015 NVIDIA Corporation. All rights reserved.
|
||||
*
|
||||
* Please refer to the NVIDIA end user license agreement (EULA) associated
|
||||
* with this source code for terms and conditions that govern your use of
|
||||
* this software. Any use, reproduction, disclosure, or distribution of
|
||||
* this software and related documentation outside the terms of the EULA
|
||||
* is strictly prohibited.
|
||||
*
|
||||
*/
|
||||
|
||||
//
|
||||
// This sample demonstrates the use of streams for concurrent execution. It also illustrates how to
|
||||
// introduce dependencies between CUDA streams with the new cudaStreamWaitEvent function introduced
|
||||
// in CUDA 3.2.
|
||||
//
|
||||
// Devices of compute capability 1.x will run the kernels one after another
|
||||
// Devices of compute capability 2.0 or higher can overlap the kernels
|
||||
//
|
||||
#include <stdio.h>
|
||||
#include <helper_functions.h>
|
||||
#include <helper_cuda.h>
|
||||
|
||||
// This is a kernel that does no real work but runs at least for a specified number of clocks
|
||||
__global__ void clock_block(clock_t *d_o, clock_t clock_count)
|
||||
{
|
||||
unsigned int start_clock = (unsigned int) clock();
|
||||
|
||||
clock_t clock_offset = 0;
|
||||
|
||||
while (clock_offset < clock_count)
|
||||
{
|
||||
unsigned int end_clock = (unsigned int) clock();
|
||||
|
||||
// The code below should work like
|
||||
// this (thanks to modular arithmetics):
|
||||
//
|
||||
// clock_offset = (clock_t) (end_clock > start_clock ?
|
||||
// end_clock - start_clock :
|
||||
// end_clock + (0xffffffffu - start_clock));
|
||||
//
|
||||
// Indeed, let m = 2^32 then
|
||||
// end - start = end + m - start (mod m).
|
||||
|
||||
clock_offset = (clock_t)(end_clock - start_clock);
|
||||
}
|
||||
|
||||
d_o[0] = clock_offset;
|
||||
}
|
||||
|
||||
|
||||
// Single warp reduction kernel
|
||||
__global__ void sum(clock_t *d_clocks, int N)
|
||||
{
|
||||
__shared__ clock_t s_clocks[32];
|
||||
|
||||
clock_t my_sum = 0;
|
||||
|
||||
for (int i = threadIdx.x; i < N; i+= blockDim.x)
|
||||
{
|
||||
my_sum += d_clocks[i];
|
||||
}
|
||||
|
||||
s_clocks[threadIdx.x] = my_sum;
|
||||
syncthreads();
|
||||
|
||||
for (int i=16; i>0; i/=2)
|
||||
{
|
||||
if (threadIdx.x < i)
|
||||
{
|
||||
s_clocks[threadIdx.x] += s_clocks[threadIdx.x + i];
|
||||
}
|
||||
|
||||
syncthreads();
|
||||
}
|
||||
|
||||
d_clocks[0] = s_clocks[0];
|
||||
}
|
||||
|
||||
int main(int argc, char **argv)
|
||||
{
|
||||
int nkernels = 8; // number of concurrent kernels
|
||||
int nstreams = nkernels + 1; // use one more stream than concurrent kernel
|
||||
int nbytes = nkernels * sizeof(clock_t); // number of data bytes
|
||||
float kernel_time = 10; // time the kernel should run in ms
|
||||
float elapsed_time; // timing variables
|
||||
int cuda_device = 0;
|
||||
|
||||
printf("[%s] - Starting...\n", argv[0]);
|
||||
|
||||
// get number of kernels if overridden on the command line
|
||||
if (checkCmdLineFlag(argc, (const char **)argv, "nkernels"))
|
||||
{
|
||||
nkernels = getCmdLineArgumentInt(argc, (const char **)argv, "nkernels");
|
||||
nstreams = nkernels + 1;
|
||||
}
|
||||
|
||||
// use command-line specified CUDA device, otherwise use device with highest Gflops/s
|
||||
cuda_device = findCudaDevice(argc, (const char **)argv);
|
||||
|
||||
// CHECK: hipDeviceProp_t deviceProp;
|
||||
cudaDeviceProp deviceProp;
|
||||
// CHECK: checkCudaErrors(hipGetDevice(&cuda_device));
|
||||
checkCudaErrors(cudaGetDevice(&cuda_device));
|
||||
|
||||
// CHECK: checkCudaErrors(hipGetDeviceProperties(&deviceProp, cuda_device));
|
||||
checkCudaErrors(cudaGetDeviceProperties(&deviceProp, cuda_device));
|
||||
|
||||
if ((deviceProp.concurrentKernels == 0))
|
||||
{
|
||||
printf("> GPU does not support concurrent kernel execution\n");
|
||||
printf(" CUDA kernel runs will be serialized\n");
|
||||
}
|
||||
|
||||
printf("> Detected Compute SM %d.%d hardware with %d multi-processors\n",
|
||||
deviceProp.major, deviceProp.minor, deviceProp.multiProcessorCount);
|
||||
|
||||
// allocate host memory
|
||||
clock_t *a = 0; // pointer to the array data in host memory
|
||||
// CHECK: checkCudaErrors(hipHostMalloc((void **)&a, nbytes));
|
||||
checkCudaErrors(cudaMallocHost((void **)&a, nbytes));
|
||||
|
||||
// allocate device memory
|
||||
clock_t *d_a = 0; // pointers to data and init value in the device memory
|
||||
// CHECK: checkCudaErrors(hipMalloc((void **)&d_a, nbytes));
|
||||
checkCudaErrors(cudaMalloc((void **)&d_a, nbytes));
|
||||
|
||||
// CHECK: hipStream_t *streams = (hipStream_t *) malloc(nstreams * sizeof(hipStream_t));
|
||||
// allocate and initialize an array of stream handles
|
||||
cudaStream_t *streams = (cudaStream_t *) malloc(nstreams * sizeof(cudaStream_t));
|
||||
|
||||
for (int i = 0; i < nstreams; i++)
|
||||
{
|
||||
// CHECK: checkCudaErrors(hipStreamCreate(&(streams[i])));
|
||||
checkCudaErrors(cudaStreamCreate(&(streams[i])));
|
||||
}
|
||||
|
||||
// CHECK: hipEvent_t start_event, stop_event;
|
||||
// create CUDA event handles
|
||||
cudaEvent_t start_event, stop_event;
|
||||
|
||||
// CHECK: checkCudaErrors(hipEventCreate(&start_event));
|
||||
// CHECK: checkCudaErrors(hipEventCreate(&stop_event));
|
||||
checkCudaErrors(cudaEventCreate(&start_event));
|
||||
checkCudaErrors(cudaEventCreate(&stop_event));
|
||||
|
||||
// the events are used for synchronization only and hence do not need to record timings
|
||||
// this also makes events not introduce global sync points when recorded which is critical to get overlap
|
||||
|
||||
// CHECK: hipEvent_t *kernelEvent;
|
||||
// CHECK: kernelEvent = (hipEvent_t *) malloc(nkernels * sizeof(hipEvent_t));
|
||||
cudaEvent_t *kernelEvent;
|
||||
kernelEvent = (cudaEvent_t *) malloc(nkernels * sizeof(cudaEvent_t));
|
||||
|
||||
for (int i = 0; i < nkernels; i++)
|
||||
{
|
||||
// CHECK: checkCudaErrors(hipEventCreateWithFlags(&(kernelEvent[i]), hipEventDisableTiming));
|
||||
checkCudaErrors(cudaEventCreateWithFlags(&(kernelEvent[i]), cudaEventDisableTiming));
|
||||
}
|
||||
|
||||
//////////////////////////////////////////////////////////////////////
|
||||
// time execution with nkernels streams
|
||||
clock_t total_clocks = 0;
|
||||
#if defined(__arm__) || defined(__aarch64__)
|
||||
// the kernel takes more time than the channel reset time on arm archs, so to prevent hangs reduce time_clocks.
|
||||
clock_t time_clocks = (clock_t)(kernel_time * (deviceProp.clockRate / 1000));
|
||||
#else
|
||||
clock_t time_clocks = (clock_t)(kernel_time * deviceProp.clockRate);
|
||||
#endif
|
||||
|
||||
// CHECK: hipEventRecord(start_event, 0);
|
||||
cudaEventRecord(start_event, 0);
|
||||
|
||||
// queue nkernels in separate streams and record when they are done
|
||||
for (int i=0; i<nkernels; ++i)
|
||||
{
|
||||
// CHECK: hipLaunchKernelGGL(clock_block, dim3(1), dim3(1), 0, streams[i], &d_a[i], time_clocks);
|
||||
clock_block<<<1,1,0,streams[i]>>>(&d_a[i], time_clocks);
|
||||
total_clocks += time_clocks;
|
||||
|
||||
// CHECK: checkCudaErrors(hipEventRecord(kernelEvent[i], streams[i]));
|
||||
checkCudaErrors(cudaEventRecord(kernelEvent[i], streams[i]));
|
||||
|
||||
// make the last stream wait for the kernel event to be recorded
|
||||
// CHECK: checkCudaErrors(hipStreamWaitEvent(streams[nstreams-1], kernelEvent[i],0));
|
||||
checkCudaErrors(cudaStreamWaitEvent(streams[nstreams-1], kernelEvent[i],0));
|
||||
}
|
||||
|
||||
// queue a sum kernel and a copy back to host in the last stream.
|
||||
// the commands in this stream get dispatched as soon as all the kernel events have been recorded
|
||||
// CHECK: hipLaunchKernelGGL(sum, dim3(1), dim3(32), 0, streams[nstreams-1], d_a, nkernels);
|
||||
// CHECK: checkCudaErrors(hipMemcpyAsync(a, d_a, sizeof(clock_t), hipMemcpyDeviceToHost, streams[nstreams-1]));
|
||||
sum<<<1,32,0,streams[nstreams-1]>>>(d_a, nkernels);
|
||||
checkCudaErrors(cudaMemcpyAsync(a, d_a, sizeof(clock_t), cudaMemcpyDeviceToHost, streams[nstreams-1]));
|
||||
|
||||
// at this point the CPU has dispatched all work for the GPU and can continue processing other tasks in parallel
|
||||
|
||||
// in this sample we just wait until the GPU is done
|
||||
// CHECK: checkCudaErrors(hipEventRecord(stop_event, 0));
|
||||
// CHECK: checkCudaErrors(hipEventSynchronize(stop_event));
|
||||
// CHECK: checkCudaErrors(hipEventElapsedTime(&elapsed_time, start_event, stop_event));
|
||||
checkCudaErrors(cudaEventRecord(stop_event, 0));
|
||||
checkCudaErrors(cudaEventSynchronize(stop_event));
|
||||
checkCudaErrors(cudaEventElapsedTime(&elapsed_time, start_event, stop_event));
|
||||
|
||||
printf("Expected time for serial execution of %d kernels = %.3fs\n", nkernels, nkernels * kernel_time/1000.0f);
|
||||
printf("Expected time for concurrent execution of %d kernels = %.3fs\n", nkernels, kernel_time/1000.0f);
|
||||
printf("Measured time for sample = %.3fs\n", elapsed_time/1000.0f);
|
||||
|
||||
bool bTestResult = (a[0] > total_clocks);
|
||||
|
||||
// release resources
|
||||
for (int i = 0; i < nkernels; i++)
|
||||
{
|
||||
// CHECK: hipStreamDestroy(streams[i]);
|
||||
// CHECK: hipEventDestroy(kernelEvent[i]);
|
||||
cudaStreamDestroy(streams[i]);
|
||||
cudaEventDestroy(kernelEvent[i]);
|
||||
}
|
||||
|
||||
free(streams);
|
||||
free(kernelEvent);
|
||||
|
||||
// CHECK: hipEventDestroy(start_event);
|
||||
// CHECK: hipEventDestroy(stop_event);
|
||||
// CHECK: hipHostFree(a);
|
||||
// CHECK: hipFree(d_a);
|
||||
cudaEventDestroy(start_event);
|
||||
cudaEventDestroy(stop_event);
|
||||
cudaFreeHost(a);
|
||||
cudaFree(d_a);
|
||||
|
||||
if (!bTestResult)
|
||||
{
|
||||
printf("Test failed!\n");
|
||||
exit(EXIT_FAILURE);
|
||||
}
|
||||
|
||||
printf("Test passed\n");
|
||||
exit(EXIT_SUCCESS);
|
||||
}
|
||||
@@ -22,7 +22,12 @@ THE SOFTWARE.
|
||||
#include<cuda.h>
|
||||
#include<cuda_runtime.h>
|
||||
#include<iostream>
|
||||
#include<unistd.h>
|
||||
#ifdef _WIN32
|
||||
#include <windows.h>
|
||||
#define sleep(x) Sleep(x)
|
||||
#else
|
||||
#include <unistd.h>
|
||||
#endif
|
||||
#include<stdio.h>
|
||||
#include<malloc.h>
|
||||
|
||||
@@ -33,12 +38,11 @@ THE SOFTWARE.
|
||||
// CHECK: if(status != hipSuccess) {
|
||||
#define check(msg, status){ \
|
||||
if(status != cudaSuccess) { \
|
||||
printf("%s failed. \n", #msg); \
|
||||
printf("%s failed. \n", #msg); \
|
||||
} \
|
||||
}
|
||||
|
||||
__global__ void Inc1(float *Ad, float *Bd){
|
||||
// CHECK: int tx = hipThreadIdx_x + hipBlockIdx_x * hipBlockDim_x;
|
||||
int tx = threadIdx.x + blockIdx.x * blockDim.x;
|
||||
if(tx < 1 ){
|
||||
for(int i=0;i<ITER;i++){
|
||||
@@ -51,7 +55,6 @@ __global__ void Inc1(float *Ad, float *Bd){
|
||||
}
|
||||
|
||||
__global__ void Inc2(float *Ad, float *Bd){
|
||||
// CHECK: int tx = hipThreadIdx_x + hipBlockIdx_x * hipBlockDim_x;
|
||||
int tx = threadIdx.x + blockIdx.x * blockDim.x;
|
||||
if(tx < 1024){
|
||||
for(int i=0;i<ITER;i++){
|
||||
|
||||
@@ -1,6 +1,8 @@
|
||||
// RUN: %run_test hipify "%s" "%t" %cuda_args
|
||||
|
||||
// CHECK: #include <hip/hip_runtime.h>
|
||||
// CHECK-NOT: #include <cuda_runtime.h>
|
||||
// CHECK: #include <stdio.h>
|
||||
#include <cuda.h>
|
||||
// CHECK-NOT: #include<cuda_runtime.h>
|
||||
#include <cuda_runtime.h>
|
||||
#include <stdio.h>
|
||||
|
||||
@@ -1,6 +1,8 @@
|
||||
// RUN: %run_test hipify "%s" "%t" %cuda_args
|
||||
|
||||
// CHECK: #include <hip/hip_runtime.h>
|
||||
#include <cuda_runtime.h>
|
||||
// CHECK-NOT: #include<cuda.h>
|
||||
#include <cuda.h>
|
||||
// CHECK: #include "hip/hip_runtime.h"
|
||||
// CHECK-NOT: #include "cuda_runtime.h"
|
||||
// CHECK: #include <stdio.h>
|
||||
#include "cuda.h"
|
||||
#include "cuda_runtime.h"
|
||||
#include <stdio.h>
|
||||
|
||||
@@ -0,0 +1,10 @@
|
||||
// RUN: %run_test hipify "%s" "%t" %cuda_args
|
||||
|
||||
// CHECK: #pragma once
|
||||
// CHECK-NEXT: #include <hip/hip_runtime.h>
|
||||
#pragma once
|
||||
// CHECK-NOT: #include <hip/hip_runtime.h>
|
||||
int main(int argc, char* argv[]) {
|
||||
return 0;
|
||||
}
|
||||
|
||||
@@ -0,0 +1,12 @@
|
||||
// RUN: %run_test hipify "%s" "%t" %cuda_args
|
||||
|
||||
// CHECK: #include <hip/hip_runtime.h>
|
||||
// CHECK-NEXT: #include <stdio.h>
|
||||
// CHECK-NEXT: #include <iostream>
|
||||
#include <stdio.h>
|
||||
#include <iostream>
|
||||
// CHECK-NOT: #include <hip/hip_runtime.h>
|
||||
int main(int argc, char* argv[]) {
|
||||
return 0;
|
||||
}
|
||||
|
||||
@@ -0,0 +1,12 @@
|
||||
// RUN: %run_test hipify "%s" "%t" %cuda_args
|
||||
|
||||
// CHECK: #pragma once
|
||||
// CHECK-NEXT: #include <hip/hip_runtime.h>
|
||||
#pragma once
|
||||
// CHECK-NOT: #include <hip/hip_runtime.h>
|
||||
#include <stdio.h>
|
||||
|
||||
int main(int argc, char* argv[]) {
|
||||
return 0;
|
||||
}
|
||||
|
||||
@@ -0,0 +1,8 @@
|
||||
// RUN: %run_test hipify "%s" "%t" %cuda_args
|
||||
|
||||
// CHECK: #include <hipblas.h>
|
||||
// CHECK-NOT: #include <cublas_v2.h>
|
||||
// CHECK: #include <stdio.h>
|
||||
#include <cublas.h>
|
||||
#include <cublas_v2.h>
|
||||
#include <stdio.h>
|
||||
@@ -0,0 +1,8 @@
|
||||
// RUN: %run_test hipify "%s" "%t" %cuda_args
|
||||
|
||||
// CHECK: #include "hipblas.h"
|
||||
// CHECK-NOT: #include "cublas.h"
|
||||
// CHECK: #include <stdio.h>
|
||||
#include "cublas_v2.h"
|
||||
#include "cublas.h"
|
||||
#include <stdio.h>
|
||||
@@ -0,0 +1,14 @@
|
||||
// RUN: %run_test hipify "%s" "%t" %cuda_args
|
||||
|
||||
// CHECK: #include <hip/hip_runtime.h>
|
||||
// CHECK-NOT: #include <cuda_runtime.h>
|
||||
// CHECK: #include <iostream>
|
||||
// CHECK: #include "hipblas.h"
|
||||
// CHECK-NOT: #include "cublas.h"
|
||||
// CHECK: #include <stdio.h>
|
||||
#include <cuda.h>
|
||||
#include <cuda_runtime.h>
|
||||
#include <iostream>
|
||||
#include "cublas_v2.h"
|
||||
#include "cublas.h"
|
||||
#include <stdio.h>
|
||||
@@ -13,7 +13,7 @@ import lit.util
|
||||
config.name = 'hipify'
|
||||
|
||||
# suffixes: CUDA source is only supported
|
||||
config.suffixes = ['.cu']
|
||||
config.suffixes = ['.cu','.cuh','.cpp','.c','.hpp','.h']
|
||||
|
||||
# testFormat: The test format to use to interpret tests.
|
||||
config.test_format = lit.formats.ShTest()
|
||||
@@ -44,8 +44,17 @@ if obj_root is not None:
|
||||
path = os.path.pathsep.join((llvm_tools_dir, config.environment['PATH']))
|
||||
config.environment['PATH'] = path
|
||||
|
||||
config.substitutions.append(("hipify", obj_root+"/hipify-clang"))
|
||||
hipify_path = obj_root
|
||||
clang_args = "-x cuda -v --cuda-gpu-arch=sm_30 --cuda-path='%s'"
|
||||
|
||||
# Clang args for CUDA...
|
||||
config.substitutions.append(("%cuda_args", "-x cuda --cuda-path=%s --cuda-gpu-arch=sm_30 -isystem%s/samples/common/inc" % (config.cuda_root, config.cuda_root)))
|
||||
config.substitutions.append(("%run_test", config.test_source_root + "/run_test.sh"))
|
||||
if sys.platform in ['win32']:
|
||||
run_test_ext = ".bat"
|
||||
hipify_path += "/" + config.build_type
|
||||
clang_args += " -isystem'%s'/common/inc -std=c++14"
|
||||
else:
|
||||
run_test_ext = ".sh"
|
||||
clang_args += " -isystem'%s'/samples/common/inc"
|
||||
|
||||
config.substitutions.append(("%cuda_args", clang_args % (config.cuda_root, config.cuda_sdk_root)))
|
||||
config.substitutions.append(("hipify", '"' + hipify_path + "/hipify-clang" + '"'))
|
||||
config.substitutions.append(("%run_test", '"' + config.test_source_root + "/run_test" + run_test_ext + '"'))
|
||||
|
||||
@@ -1,8 +1,23 @@
|
||||
import sys
|
||||
import os
|
||||
|
||||
config.llvm_tools_dir = "@LLVM_TOOLS_BINARY_DIR@"
|
||||
config.obj_root = "@CMAKE_CURRENT_BINARY_DIR@"
|
||||
config.cuda_root = "@CUDA_TOOLKIT_ROOT_DIR@"
|
||||
if sys.platform in ['win32']:
|
||||
config.cuda_sdk_root = "@CUDA_SDK_ROOT_DIR@"
|
||||
if not config.cuda_sdk_root or config.cuda_sdk_root == "CUDA_SDK_ROOT_DIR-NOTFOUND":
|
||||
cuda_version = "@CUDA_VERSION@"
|
||||
cuda_version = cuda_version.replace('.','_')
|
||||
config.cuda_samples_root = os.environ.get('NVCUDASAMPLES' + cuda_version + '_ROOT')
|
||||
if not config.cuda_samples_root:
|
||||
lit_config.fatal('No CUDA Samples dir set! Please set CUDA_SDK_ROOT_DIR.')
|
||||
config.cuda_sdk_root = config.cuda_samples_root
|
||||
config.build_type = "@CMAKE_BUILD_TYPE@"
|
||||
if not config.build_type:
|
||||
config.build_type = "Debug"
|
||||
else:
|
||||
config.cuda_sdk_root = config.cuda_root
|
||||
|
||||
# Support substitution of the tools and libs dirs with user parameters. This is
|
||||
# used when we can't determine the tool dir at configuration time.
|
||||
|
||||
@@ -0,0 +1,19 @@
|
||||
@echo off
|
||||
setlocal
|
||||
|
||||
for %%i in (FileCheck.exe) do set FILE_CHECK=%%~$PATH:i
|
||||
if not defined FILE_CHECK (echo Error: FileCheck.exe not found in PATH. && exit /b 1)
|
||||
|
||||
set HIPIFY=%1
|
||||
set IN_FILE=%2
|
||||
set TMP_FILE=%3
|
||||
|
||||
set all_args=%*
|
||||
call set clang_args=%%all_args:*%4=%%
|
||||
set clang_args=%4%clang_args%
|
||||
|
||||
%HIPIFY% -o=%TMP_FILE% %IN_FILE% -- %clang_args%
|
||||
if errorlevel 1 (echo Error: hipify-clang.exe failed with exit code: %errorlevel% && exit /b %errorlevel%)
|
||||
|
||||
findstr /v /r /c:"[ ]*//[ ]*[CHECK*|RUN]" %TMP_FILE% | %FILE_CHECK% %IN_FILE%
|
||||
if errorlevel 1 (echo Error: FileCheck.exe failed with exit code: %errorlevel% && exit /b %errorlevel%)
|
||||
@@ -13,16 +13,5 @@ shift 3
|
||||
|
||||
# Remaining args are the ones to forward to clang proper.
|
||||
|
||||
# Time for the classic insane little trick for making colour output work.
|
||||
# A self-deleting shell-script that does the thing we want to do...
|
||||
TMP_SCRIPT=$(mktemp)
|
||||
cat << EOF > $TMP_SCRIPT
|
||||
set -o errexit
|
||||
set -o xtrace
|
||||
rm $TMP_SCRIPT
|
||||
$HIPIFY -o=$TMP_FILE $IN_FILE -- $@ && cat $TMP_FILE | sed -Ee 's|//.+|// |g' | FileCheck $IN_FILE
|
||||
EOF
|
||||
chmod a+x $TMP_SCRIPT
|
||||
|
||||
# Run the script via socat, spawning a virtual terminal and propagating exit code, and hence failure.
|
||||
socat -du EXEC:$TMP_SCRIPT,pty,stderr STDOUT
|
||||
|
||||
@@ -41,8 +41,6 @@ template <typename T>
|
||||
__global__ void
|
||||
vector_square(T *C_d, const T *A_d, size_t N)
|
||||
{
|
||||
// CHECK: size_t offset = (hipBlockIdx_x * hipBlockDim_x + hipThreadIdx_x);
|
||||
// CHECK: size_t stride = hipBlockDim_x * hipGridDim_x;
|
||||
size_t offset = (blockIdx.x * blockDim.x + threadIdx.x);
|
||||
size_t stride = blockDim.x * gridDim.x;
|
||||
|
||||
|
||||
@@ -21,7 +21,7 @@ THE SOFTWARE.
|
||||
*/
|
||||
|
||||
/* HIT_START
|
||||
* BUILD: %t %s ../test_common.cpp
|
||||
* BUILD: %t %s ../test_common.cpp NVCC_OPTIONS --Wno-deprecated-declarations
|
||||
* RUN: %t
|
||||
* HIT_END
|
||||
*/
|
||||
|
||||
@@ -18,7 +18,7 @@ THE SOFTWARE.
|
||||
*/
|
||||
|
||||
/* HIT_START
|
||||
* BUILD: %t %s ../test_common.cpp
|
||||
* BUILD: %t %s ../test_common.cpp NVCC_OPTIONS --Wno-deprecated-declarations
|
||||
* RUN: %t
|
||||
* HIT_END
|
||||
*/
|
||||
|
||||
@@ -397,32 +397,30 @@ int main(int argc, char *argv[])
|
||||
if (gpuCount < 2)
|
||||
{
|
||||
printf("P2P application requires atleast 2 gpu devices\n");
|
||||
return 0;
|
||||
}
|
||||
} else {
|
||||
if (p_tests & 0x100) {
|
||||
testPeerHostToDevice(false/*useAsyncCopy*/);
|
||||
}
|
||||
testPeerHostToDevice(true/*useAsyncCopy*/);
|
||||
|
||||
if (p_tests & 0x100) {
|
||||
testPeerHostToDevice(false/*useAsyncCopy*/);
|
||||
}
|
||||
testPeerHostToDevice(true/*useAsyncCopy*/);
|
||||
if (p_tests & 0x1) {
|
||||
enablePeerFirst(false/*useAsyncCopy*/);
|
||||
}
|
||||
|
||||
if (p_tests & 0x1) {
|
||||
enablePeerFirst(false/*useAsyncCopy*/);
|
||||
}
|
||||
if (p_tests & 0x2) {
|
||||
allocMemoryFirst(false/*useAsyncCopy*/);
|
||||
}
|
||||
|
||||
if (p_tests & 0x2) {
|
||||
allocMemoryFirst(false/*useAsyncCopy*/);
|
||||
}
|
||||
if (p_tests & 0x4) {
|
||||
simpleNegative();
|
||||
}
|
||||
|
||||
if (p_tests & 0x4) {
|
||||
simpleNegative();
|
||||
if (p_tests & 0x8) {
|
||||
enablePeerFirst(true/*useAsyncCopy*/);
|
||||
}
|
||||
if (p_tests & 0x10) {
|
||||
allocMemoryFirst(true/*useAsyncCopy*/);
|
||||
}
|
||||
}
|
||||
|
||||
if (p_tests & 0x8) {
|
||||
enablePeerFirst(true/*useAsyncCopy*/);
|
||||
}
|
||||
if (p_tests & 0x10) {
|
||||
allocMemoryFirst(true/*useAsyncCopy*/);
|
||||
}
|
||||
|
||||
passed();
|
||||
}
|
||||
|
||||
@@ -23,7 +23,7 @@ THE SOFTWARE.
|
||||
|
||||
/* HIT_START
|
||||
* BUILD: %t %s ../../test_common.cpp
|
||||
* RUN: %t EXCLUDE_HIP_PLATFORM nvcc
|
||||
* RUN: %t
|
||||
* HIT_END
|
||||
*/
|
||||
|
||||
|
||||
Посилання в новій задачі
Заблокувати користувача