From 4094b624079b793fa959a22ea91755cc097908e3 Mon Sep 17 00:00:00 2001 From: Jeff Daily Date: Wed, 10 Jul 2019 20:11:41 +0000 Subject: [PATCH 01/11] remove stream locks where it is safe to do so [ROCm/hip commit: f096a3239ed3c813c8f04e829bcfac0d99e2beda] --- projects/hip/src/hip_event.cpp | 4 +--- projects/hip/src/hip_hcc.cpp | 30 ++++++++++------------------- projects/hip/src/hip_hcc_internal.h | 4 +--- projects/hip/src/hip_stream.cpp | 5 ++--- 4 files changed, 14 insertions(+), 29 deletions(-) diff --git a/projects/hip/src/hip_event.cpp b/projects/hip/src/hip_event.cpp index 7438c7d3d0..73e296dae3 100644 --- a/projects/hip/src/hip_event.cpp +++ b/projects/hip/src/hip_event.cpp @@ -191,10 +191,8 @@ hipError_t hipEventSynchronize(hipEvent_t event) { ctx->locked_syncDefaultStream(true, true); return ihipLogStatus(hipSuccess); } else { - ecd._stream->locked_eventWaitComplete( - ecd.marker(), (event->_flags & hipEventBlockingSync) ? hc::hcWaitModeBlocked + ecd.marker().wait((event->_flags & hipEventBlockingSync) ? hc::hcWaitModeBlocked : hc::hcWaitModeActive); - return ihipLogStatus(hipSuccess); } } else { diff --git a/projects/hip/src/hip_hcc.cpp b/projects/hip/src/hip_hcc.cpp index 1b11b3c67e..1b2f895f7d 100644 --- a/projects/hip/src/hip_hcc.cpp +++ b/projects/hip/src/hip_hcc.cpp @@ -316,16 +316,20 @@ void ihipStream_t::wait(LockedAccessor_StreamCrit_t& crit) { tprintf(DB_SYNC, "%s wait for queue-empty..\n", ToString(this).c_str()); crit->_av.wait(waitMode()); - - crit->_kernelCnt = 0; } //--- // Wait for all kernel and data copy commands in this stream to complete. void ihipStream_t::locked_wait() { - LockedAccessor_StreamCrit_t crit(_criticalData); + // create a marker while holding stream lock, + // but release lock prior to waiting on the marker + hc::completion_future marker; + { + LockedAccessor_StreamCrit_t crit(_criticalData); + marker = crit->_av.create_marker(hc::no_scope); + } - wait(crit); + marker.wait(waitMode()); }; // Causes current stream to wait for specified event to complete: @@ -340,30 +344,14 @@ void ihipStream_t::locked_streamWaitEvent(ihipEventData_t& ecd) { // Causes current stream to wait for specified event to complete: // Note this does not provide any kind of host serialization. bool ihipStream_t::locked_eventIsReady(hipEvent_t event) { - // Event query that returns "Complete" may cause HCC to manipulate - // internal queue state so lock the stream's queue here. - LockedAccessor_StreamCrit_t scrit(_criticalData); - LockedAccessor_EventCrit_t ecrit(event->criticalData()); return (ecrit->_eventData.marker().is_ready()); } -// Waiting on event can cause HCC to reclaim stream resources - so need to lock the stream. -void ihipStream_t::locked_eventWaitComplete(hc::completion_future& marker, - hc::hcWaitMode waitMode) { - LockedAccessor_StreamCrit_t crit(_criticalData); - - marker.wait(waitMode); -} - - // Create a marker in this stream. // Save state in the event so it can track the status of the event. hc::completion_future ihipStream_t::locked_recordEvent(hipEvent_t event) { - // Lock the stream to prevent simultaneous access - LockedAccessor_StreamCrit_t crit(_criticalData); - auto scopeFlag = hc::accelerator_scope; // The env var HIP_EVENT_SYS_RELEASE sets the default, // The explicit flags override the env var (if specified) @@ -375,6 +363,8 @@ hc::completion_future ihipStream_t::locked_recordEvent(hipEvent_t event) { scopeFlag = HIP_EVENT_SYS_RELEASE ? hc::system_scope : hc::accelerator_scope; } + // Lock the stream to prevent simultaneous access + LockedAccessor_StreamCrit_t crit(_criticalData); return crit->_av.create_marker(scopeFlag); }; diff --git a/projects/hip/src/hip_hcc_internal.h b/projects/hip/src/hip_hcc_internal.h index a07ff0e7e7..7ab3fa6de1 100644 --- a/projects/hip/src/hip_hcc_internal.h +++ b/projects/hip/src/hip_hcc_internal.h @@ -476,7 +476,7 @@ template class ihipStreamCriticalBase_t : public LockedBase { public: ihipStreamCriticalBase_t(ihipStream_t* parentStream, hc::accelerator_view av) - : _kernelCnt(0), _av(av), _parent(parentStream){}; + : _av(av), _parent(parentStream){}; ~ihipStreamCriticalBase_t() {} @@ -500,7 +500,6 @@ class ihipStreamCriticalBase_t : public LockedBase { public: ihipStream_t* _parent; - uint32_t _kernelCnt; // Count of inflight kernels in this stream. Reset at ::wait(). hc::accelerator_view _av; @@ -564,7 +563,6 @@ class ihipStream_t { hc::completion_future locked_recordEvent(hipEvent_t event); bool locked_eventIsReady(hipEvent_t event); - void locked_eventWaitComplete(hc::completion_future& marker, hc::hcWaitMode waitMode); ihipStreamCritical_t& criticalData() { return _criticalData; }; diff --git a/projects/hip/src/hip_stream.cpp b/projects/hip/src/hip_stream.cpp index d70f860f90..7812530dda 100644 --- a/projects/hip/src/hip_stream.cpp +++ b/projects/hip/src/hip_stream.cpp @@ -142,9 +142,8 @@ hipError_t hipStreamWaitEvent(hipStream_t stream, hipEvent_t event, unsigned int // conservative wait on host for the specified event to complete: // return _stream->locked_eventWaitComplete(this, waitMode); // - ecd._stream->locked_eventWaitComplete( - ecd.marker(), (event->_flags & hipEventBlockingSync) ? hc::hcWaitModeBlocked - : hc::hcWaitModeActive); + ecd.marker().wait((event->_flags & hipEventBlockingSync) ? hc::hcWaitModeBlocked + : hc::hcWaitModeActive); } else { stream = ihipSyncAndResolveStream(stream); // This will use create_blocking_marker to wait on the specified queue. From a53639b073281e2e5b23343a41a62bdc52ffb9ed Mon Sep 17 00:00:00 2001 From: Rahul Garg Date: Tue, 30 Jul 2019 02:56:47 +0530 Subject: [PATCH 02/11] [docs]Fix texture reference APIs usage part [ROCm/hip commit: ccaea193b22d1b12c7d306c32303891d872e171a] --- .../hip/docs/markdown/hip_porting_guide.md | 34 +------------------ 1 file changed, 1 insertion(+), 33 deletions(-) diff --git a/projects/hip/docs/markdown/hip_porting_guide.md b/projects/hip/docs/markdown/hip_porting_guide.md index 90854c3a46..6915945d4e 100644 --- a/projects/hip/docs/markdown/hip_porting_guide.md +++ b/projects/hip/docs/markdown/hip_porting_guide.md @@ -485,39 +485,7 @@ AMD compilers currently load all data into both the L1 and L2 caches, so __ldg i We recommend the following for functional portability: - For programs that use textures only to benefit from improved caching, use the __ldg instruction -- Programs that use texture object APIs, work well on HIP -- For program that use texture reference APIs, use conditional compilation (see [Identify HIP Target Platform](#identify-hip-target-platform)) - - For the `__HIP_PLATFORM_HCC__` path, pass an additional argument to the kernel and in texture fetch API inside kernel as shown below:- - -``` -texture tex; - -__global__ void tex2DKernel(float* outputData, -#ifdef __HIP_PLATFORM_HCC__ - hipTextureObject_t textureObject, -#endif - int width, - int height) -{ - int x = blockIdx.x*blockDim.x + threadIdx.x; - int y = blockIdx.y*blockDim.y + threadIdx.y; -#ifdef __HIP_PLATFORM_HCC__ - outputData[y*width + x] = tex2D(tex, textureObject, x, y); -#else - outputData[y*width + x] = tex2D(tex, x, y); -#endif -} - -// Host code: -void myFunc () -{ - // ... - -#ifdef __HIP_PLATFORM_HCC__ - hipLaunchKernelGGL(tex2DKernel, dim3(dimGrid), dim3(dimBlock), 0, 0, dData, tex.textureObject, width, height); -#else - hipLaunchKernelGGL(tex2DKernel, dim3(dimGrid), dim3(dimBlock), 0, 0, dData, width, height); -#endif +- Programs that use texture object and reference APIs, work well on HIP ``` From 0009dc1067da0825ba7df2d6ff440183df83b0ea Mon Sep 17 00:00:00 2001 From: Rahul Garg Date: Wed, 31 Jul 2019 00:28:30 +0530 Subject: [PATCH 03/11] Change hipErrorUnknown to hipErrorInvalidValue [ROCm/hip commit: 483aab031faf02c4888b8f32829c265820745fad] --- projects/hip/src/hip_memory.cpp | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/projects/hip/src/hip_memory.cpp b/projects/hip/src/hip_memory.cpp index 2355772fd6..99ea4b5b24 100644 --- a/projects/hip/src/hip_memory.cpp +++ b/projects/hip/src/hip_memory.cpp @@ -198,7 +198,7 @@ hipError_t hipPointerGetAttributes(hipPointerAttribute_t* attributes, const void attributes->isManaged = 0; attributes->allocationFlags = 0; - e = hipErrorUnknown; // TODO - should be hipErrorInvalidValue ? + e = hipErrorInvalidValue; } } return ihipLogStatus(e); From d7973153cad429360a57da8d2cb3822884352194 Mon Sep 17 00:00:00 2001 From: Rahul Garg Date: Wed, 31 Jul 2019 08:42:08 -0700 Subject: [PATCH 04/11] Add HIP init in hipFuncGetAttributes (#1262) * Add HIP init in hipFuncGetAttributes * [dtest]Remove explicit hip init call in hipFuncGetAttributes dtest [ROCm/hip commit: 0517c305078573dfe9704d785766ab2a0a5c88b9] --- projects/hip/src/hip_module.cpp | 3 ++- .../hip/tests/src/runtimeApi/module/hipFuncGetAttributes.cpp | 1 - 2 files changed, 2 insertions(+), 2 deletions(-) diff --git a/projects/hip/src/hip_module.cpp b/projects/hip/src/hip_module.cpp index c9b6c39c29..221c3850de 100644 --- a/projects/hip/src/hip_module.cpp +++ b/projects/hip/src/hip_module.cpp @@ -785,6 +785,7 @@ hipFuncAttributes make_function_attributes(const ihipModuleSymbol_t& kd) { hipError_t hipFuncGetAttributes(hipFuncAttributes* attr, const void* func) { + HIP_INIT_API(hipFuncGetAttributes, attr, func); using namespace hip_impl; if (!attr) return hipErrorInvalidValue; @@ -797,7 +798,7 @@ hipError_t hipFuncGetAttributes(hipFuncAttributes* attr, const void* func) *attr = make_function_attributes(*kd); - return hipSuccess; + return ihipLogStatus(hipSuccess); } hipError_t ihipModuleLoadData(hipModule_t* module, const void* image) { diff --git a/projects/hip/tests/src/runtimeApi/module/hipFuncGetAttributes.cpp b/projects/hip/tests/src/runtimeApi/module/hipFuncGetAttributes.cpp index 9ed9beb078..0003c327fe 100644 --- a/projects/hip/tests/src/runtimeApi/module/hipFuncGetAttributes.cpp +++ b/projects/hip/tests/src/runtimeApi/module/hipFuncGetAttributes.cpp @@ -39,7 +39,6 @@ void fn(float* px, float* py) } int main() { - hipInit(0); hipFuncAttributes attr{}; From 617a6d43dcb464714b540263ed780d0e4fa64c87 Mon Sep 17 00:00:00 2001 From: Rahul Garg Date: Wed, 31 Jul 2019 08:42:29 -0700 Subject: [PATCH 05/11] Add hip init in hipExtLaunchMultiKernelMultiDevice (#1263) * Add hip init in hipExtLaunchMultiKernelMultiDevice * Add more logstatus for multiple return paths * Fix missing i in function name [ROCm/hip commit: b9e6d72ee6bb46fd560e967c7790a9e28967111d] --- projects/hip/src/hip_module.cpp | 12 ++++++------ 1 file changed, 6 insertions(+), 6 deletions(-) diff --git a/projects/hip/src/hip_module.cpp b/projects/hip/src/hip_module.cpp index 221c3850de..2d3d052036 100644 --- a/projects/hip/src/hip_module.cpp +++ b/projects/hip/src/hip_module.cpp @@ -318,16 +318,16 @@ hipError_t hipHccModuleLaunchKernel(hipFunction_t f, uint32_t globalWorkSizeX, hipError_t hipExtLaunchMultiKernelMultiDevice(hipLaunchParams* launchParamsList, int numDevices, unsigned int flags) { - + HIP_INIT_API(hipExtLaunchMultiKernelMultiDevice, launchParamsList,numDevices,flags); hipError_t result; if ((numDevices > g_deviceCnt) || (launchParamsList == nullptr)) { - return hipErrorInvalidValue; + return ihipLogStatus(hipErrorInvalidValue); } hipFunction_t* kds = reinterpret_cast(malloc(sizeof(hipFunction_t) * numDevices)); if (kds == nullptr) { - return hipErrorNotInitialized; + return ihipLogStatus(hipErrorNotInitialized); } // prepare all kernel descriptors for each device as all streams will be locked in the next loop @@ -335,13 +335,13 @@ hipError_t hipExtLaunchMultiKernelMultiDevice(hipLaunchParams* launchParamsList, const hipLaunchParams& lp = launchParamsList[i]; if (lp.stream == nullptr) { free(kds); - return hipErrorNotInitialized; + return ihipLogStatus(hipErrorNotInitialized); } kds[i] = hip_impl::get_program_state().kernel_descriptor(reinterpret_cast(lp.func), hip_impl::target_agent(lp.stream)); if (kds[i] == nullptr) { free(kds); - return hipErrorInvalidValue; + return ihipLogStatus(hipErrorInvalidValue); } hip_impl::kernargs_size_align kargs = hip_impl::get_program_state().get_kernargs_size_align( reinterpret_cast(lp.func)); @@ -382,7 +382,7 @@ hipError_t hipExtLaunchMultiKernelMultiDevice(hipLaunchParams* launchParamsList, free(kds); - return result; + return ihipLogStatus(result); } namespace hip_impl { From 2810475577b91fd8c8666be2be2a0cf3ff619d65 Mon Sep 17 00:00:00 2001 From: Evgeny Mankov Date: Wed, 31 Jul 2019 22:59:05 +0300 Subject: [PATCH 06/11] [HIP][doc] Populate CUDA Runtime API doc with CUDA version field + CUDA version - version in which API has appeared and (optional) last version before abandoning it; no value in case of earlier versions < 7.5. + Fix typos, add missing references. [ROCm/hip commit: 77e9ade9bc80ecb73c4dae19abc635aab84cb4b4] --- ..._Runtime_API_functions_supported_by_HIP.md | 1551 +++++++++-------- .../src/CUDA2HIP_Runtime_API_types.cpp | 5 +- 2 files changed, 781 insertions(+), 775 deletions(-) diff --git a/projects/hip/docs/markdown/CUDA_Runtime_API_functions_supported_by_HIP.md b/projects/hip/docs/markdown/CUDA_Runtime_API_functions_supported_by_HIP.md index afd05b7d87..1bca61f33b 100644 --- a/projects/hip/docs/markdown/CUDA_Runtime_API_functions_supported_by_HIP.md +++ b/projects/hip/docs/markdown/CUDA_Runtime_API_functions_supported_by_HIP.md @@ -2,39 +2,39 @@ ## **1. Device Management** -| **CUDA** | **HIP** | -|-----------------------------------------------------------|-------------------------------| -| `cudaChooseDevice` | `hipChooseDevice` | -| `cudaDeviceGetAttribute` | `hipDeviceGetAttribute` | -| `cudaDeviceGetByPCIBusId` | `hipDeviceGetByPCIBusId` | -| `cudaDeviceGetCacheConfig` | `hipDeviceGetCacheConfig` | -| `cudaDeviceGetLimit` | `hipDeviceGetLimit` | -| `cudaDeviceGetPCIBusId` | `hipDeviceGetPCIBusId` | -| `cudaDeviceGetSharedMemConfig` | `hipDeviceGetSharedMemConfig` | +| **CUDA** | **HIP** |**CUDA version\***| +|-----------------------------------------------------------|-----------------------------------|:----------------:| +| `cudaChooseDevice` | `hipChooseDevice` | +| `cudaDeviceGetAttribute` | `hipDeviceGetAttribute` | +| `cudaDeviceGetByPCIBusId` | `hipDeviceGetByPCIBusId` | +| `cudaDeviceGetCacheConfig` | `hipDeviceGetCacheConfig` | +| `cudaDeviceGetLimit` | `hipDeviceGetLimit` | +| `cudaDeviceGetPCIBusId` | `hipDeviceGetPCIBusId` | +| `cudaDeviceGetSharedMemConfig` | `hipDeviceGetSharedMemConfig` | | `cudaDeviceGetStreamPriorityRange` | `hipDeviceGetStreamPriorityRange` | -| `cudaDeviceReset` | `hipDeviceReset` | -| `cudaDeviceSetCacheConfig` | `hipDeviceSetCacheConfig` | -| `cudaDeviceSetLimit` | `hipDeviceSetLimit` | -| `cudaDeviceSetSharedMemConfig` | `hipDeviceSetSharedMemConfig` | -| `cudaDeviceSynchronize` | `hipDeviceSynchronize` | -| `cudaGetDevice` | `hipGetDevice` | -| `cudaGetDeviceCount` | `hipGetDeviceCount` | -| `cudaGetDeviceFlags` | `hipCtxGetFlags` | -| `cudaGetDeviceProperties` | `hipGetDeviceProperties` | -| `cudaIpcCloseMemHandle` | `hipIpcCloseMemHandle` | -| `cudaIpcGetEventHandle` | `hipIpcGetEventHandle` | -| `cudaIpcGetMemHandle` | `hipIpcGetMemHandle` | -| `cudaIpcOpenEventHandle` | `hipIpcOpenEventHandle` | -| `cudaIpcOpenMemHandle` | `hipIpcOpenMemHandle` | -| `cudaSetDevice` | `hipSetDevice` | -| `cudaSetDeviceFlags` | `hipSetDeviceFlags` | -| `cudaSetValidDevices` | | -| `cudaDeviceGetP2PAttribute` | | +| `cudaDeviceReset` | `hipDeviceReset` | +| `cudaDeviceSetCacheConfig` | `hipDeviceSetCacheConfig` | +| `cudaDeviceSetLimit` | `hipDeviceSetLimit` | +| `cudaDeviceSetSharedMemConfig` | `hipDeviceSetSharedMemConfig` | +| `cudaDeviceSynchronize` | `hipDeviceSynchronize` | +| `cudaGetDevice` | `hipGetDevice` | +| `cudaGetDeviceCount` | `hipGetDeviceCount` | +| `cudaGetDeviceFlags` | `hipCtxGetFlags` | +| `cudaGetDeviceProperties` | `hipGetDeviceProperties` | +| `cudaIpcCloseMemHandle` | `hipIpcCloseMemHandle` | +| `cudaIpcGetEventHandle` | `hipIpcGetEventHandle` | +| `cudaIpcGetMemHandle` | `hipIpcGetMemHandle` | +| `cudaIpcOpenEventHandle` | `hipIpcOpenEventHandle` | +| `cudaIpcOpenMemHandle` | `hipIpcOpenMemHandle` | +| `cudaSetDevice` | `hipSetDevice` | +| `cudaSetDeviceFlags` | `hipSetDeviceFlags` | +| `cudaSetValidDevices` | | +| `cudaDeviceGetP2PAttribute` | | 8.0 | ## **2. Thread Management [DEPRECATED]** -| **CUDA** | **HIP** | -|-----------------------------------------------------------|-------------------------------| +| **CUDA** | **HIP** |**CUDA version\***| +|-----------------------------------------------------------|-------------------------------|:----------------:| | `cudaThreadExit` | `hipDeviceReset` | | `cudaThreadGetCacheConfig` | `hipDeviceGetCacheConfig` | | `cudaThreadGetLimit` | | @@ -44,8 +44,8 @@ ## **3. Error Handling** -| **CUDA** | **HIP** | -|-----------------------------------------------------------|-------------------------------| +| **CUDA** | **HIP** |**CUDA version\***| +|-----------------------------------------------------------|-------------------------------|:----------------:| | `cudaGetErrorName` | `hipGetErrorName` | | `cudaGetErrorString` | `hipGetErrorString` | | `cudaGetLastError` | `hipGetLastError` | @@ -53,14 +53,14 @@ ## **4. Stream Management** -| **CUDA** | **HIP** | -|-----------------------------------------------------------|-------------------------------| +| **CUDA** | **HIP** |**CUDA version\***| +|-----------------------------------------------------------|-------------------------------|:----------------:| | `cudaStreamAddCallback` | `hipStreamAddCallback` | | `cudaStreamAttachMemAsync` | | -| `cudaStreamBeginCapture` | | -| `cudaStreamEndCapture` | | -| `cudaStreamIsCapturing` | | -| `cudaStreamGetCaptureInfo` | | +| `cudaStreamBeginCapture` | | 10.0 | +| `cudaStreamEndCapture` | | 10.0 | +| `cudaStreamIsCapturing` | | 10.0 | +| `cudaStreamGetCaptureInfo` | | 10.1 | | `cudaStreamCreate` | `hipStreamCreate` | | `cudaStreamCreateWithFlags` | `hipStreamCreateWithFlags` | | `cudaStreamCreateWithPriority` | `hipStreamCreateWithPriority` | @@ -70,12 +70,12 @@ | `cudaStreamQuery` | `hipStreamQuery` | | `cudaStreamSynchronize` | `hipStreamSynchronize` | | `cudaStreamWaitEvent` | `hipStreamWaitEvent` | -| `cudaThreadExchangeStreamCaptureMode` | | +| `cudaThreadExchangeStreamCaptureMode` | | 10.1 | ## **5. Event Management** -| **CUDA** | **HIP** | -|-----------------------------------------------------------|-------------------------------| +| **CUDA** | **HIP** |**CUDA version\***| +|-----------------------------------------------------------|-------------------------------|:----------------:| | `cudaEventCreate` | `hipEventCreate` | | `cudaEventCreateWithFlags` | `hipEventCreateWithFlags` | | `cudaEventDestroy` | `hipEventDestroy` | @@ -86,23 +86,23 @@ ## **6. External Resource Interoperability** -| **CUDA** | **HIP** | -|-----------------------------------------------------------|-------------------------------| -| `cudaSignalExternalSemaphoresAsync` | | -| `cudaWaitExternalSemaphoresAsync` | | -| `cudaImportExternalMemory` | | -| `cudaExternalMemoryGetMappedBuffer` | | -| `cudaExternalMemoryGetMappedMipmappedArray` | | -| `cudaDestroyExternalMemory` | | -| `cudaImportExternalSemaphore` | | -| `cudaDestroyExternalSemaphore` | | +| **CUDA** | **HIP** |**CUDA version\***| +|-----------------------------------------------------------|-------------------------------|:----------------:| +| `cudaSignalExternalSemaphoresAsync` | | 10.0 | +| `cudaWaitExternalSemaphoresAsync` | | 10.0 | +| `cudaImportExternalMemory` | | 10.0 | +| `cudaExternalMemoryGetMappedBuffer` | | 10.0 | +| `cudaExternalMemoryGetMappedMipmappedArray` | | 10.0 | +| `cudaDestroyExternalMemory` | | 10.0 | +| `cudaImportExternalSemaphore` | | 10.0 | +| `cudaDestroyExternalSemaphore` | | 10.0 | ## **7. Execution Control** -| **CUDA** | **HIP** | -|-----------------------------------------------------------|-------------------------------| +| **CUDA** | **HIP** |**CUDA version\***| +|-----------------------------------------------------------|-------------------------------|:----------------:| | `cudaFuncGetAttributes` | | -| `cudaFuncSetAttribute` | | +| `cudaFuncSetAttribute` | | 9.0 | | `cudaFuncSetCacheConfig` | `hipFuncSetCacheConfig` | | `cudaFuncSetSharedMemConfig` | | | `cudaGetParameterBuffer` | | @@ -110,29 +110,29 @@ | `cudaLaunchKernel` | `hipLaunchKernel` | | `cudaSetDoubleForDevice` | | | `cudaSetDoubleForHost` | | -| `cudaLaunchCooperativeKernel` | | -| `cudaLaunchCooperativeKernelMultiDevice` | | -| `cudaLaunchHostFunc` | | +| `cudaLaunchCooperativeKernel` | | 9.0 | +| `cudaLaunchCooperativeKernelMultiDevice` | | 9.0 | +| `cudaLaunchHostFunc` | | 10.0 | ## **8. Occupancy** -| **CUDA** | **HIP** | -|-----------------------------------------------------------|-----------------------------------------------| +| **CUDA** | **HIP** |**CUDA version\***| +|-----------------------------------------------------------|-----------------------------------------------|:----------------:| | `cudaOccupancyMaxActiveBlocksPerMultiprocessor` | `hipOccupancyMaxActiveBlocksPerMultiprocessor`| | `cudaOccupancyMaxActiveBlocksPerMultiprocessorWithFlags` | | ## **9. Execution Control [DEPRECATED since 7.0]** -| **CUDA** | **HIP** | -|-----------------------------------------------------------|-------------------------------| +| **CUDA** | **HIP** |**CUDA version\***| +|-----------------------------------------------------------|-------------------------------|:----------------:| | `cudaConfigureCall` | `hipConfigureCall` | | `cudaLaunch` | `hipLaunchByPtr` | -| `cudaSetupArgument` | `hipSetupArgument` | +| `cudaSetupArgument` | `hipSetupArgument` | 7.0 - 10.0 | ## **10. Memory Management** -| **CUDA** | **HIP** | -|-----------------------------------------------------------|-------------------------------| +| **CUDA** | **HIP** |**CUDA version\***| +|-----------------------------------------------------------|-------------------------------|:----------------:| | `cudaArrayGetInfo` | | | `cudaFree` | `hipFree` | | `cudaFreeArray` | `hipFreeArray` | @@ -155,7 +155,7 @@ | `cudaMallocMipmappedArray` | | | `cudaMallocPitch` | | | `cudaMemGetInfo` | `hipMemGetInfo` | -| `cudaMemPrefetchAsync` | | +| `cudaMemPrefetchAsync` | | 8.0 | | `cudaMemcpy` | `hipMemcpy` | | `cudaMemcpy2D` | `hipMemcpy2D` | | `cudaMemcpy2DArrayToArray` | | @@ -192,22 +192,22 @@ ## **11. Unified Addressing** -| **CUDA** | **HIP** | -|-----------------------------------------------------------|-------------------------------| +| **CUDA** | **HIP** |**CUDA version\***| +|-----------------------------------------------------------|-------------------------------|:----------------:| | `cudaPointerGetAttributes` | `hipPointerGetAttributes` | ## **12. Peer Device Memory Access** -| **CUDA** | **HIP** | -|-----------------------------------------------------------|-------------------------------| +| **CUDA** | **HIP** |**CUDA version\***| +|-----------------------------------------------------------|-------------------------------|:----------------:| | `cudaDeviceCanAccessPeer` | `hipDeviceCanAccessPeer` | | `cudaDeviceDisablePeerAccess` | `hipDeviceDisablePeerAccess` | | `cudaDeviceEnablePeerAccess` | `hipDeviceEnablePeerAccess` | ## **13. OpenGL Interoperability** -| **CUDA** | **HIP** | -|-----------------------------------------------------------|-------------------------------| +| **CUDA** | **HIP** |**CUDA version\***| +|-----------------------------------------------------------|-------------------------------|:----------------:| | `cudaGLGetDevices` | | | `cudaGraphicsGLRegisterBuffer` | | | `cudaGraphicsGLRegisterImage` | | @@ -215,8 +215,8 @@ ## **14. OpenGL Interoperability [DEPRECATED]** -| **CUDA** | **HIP** | -|-----------------------------------------------------------|-------------------------------| +| **CUDA** | **HIP** |**CUDA version\***| +|-----------------------------------------------------------|-------------------------------|:----------------:| | `cudaGLMapBufferObject` | | | `cudaGLMapBufferObjectAsync` | | | `cudaGLRegisterBufferObject` | | @@ -228,8 +228,8 @@ ## **15. Direct3D 9 Interoperability** -| **CUDA** | **HIP** | -|-----------------------------------------------------------|-------------------------------| +| **CUDA** | **HIP** |**CUDA version\***| +|-----------------------------------------------------------|-------------------------------|:----------------:| | `cudaD3D9GetDevice` | | | `cudaD3D9GetDevices` | | | `cudaD3D9GetDirect3DDevice` | | @@ -238,8 +238,8 @@ ## **16. Direct3D 9 Interoperability [DEPRECATED]** -| **CUDA** | **HIP** | -|-----------------------------------------------------------|-------------------------------| +| **CUDA** | **HIP** |**CUDA version\***| +|-----------------------------------------------------------|-------------------------------|:----------------:| | `cudaD3D9MapResources` | | | `cudaD3D9RegisterResource` | | | `cudaD3D9ResourceGetMappedArray` | | @@ -253,16 +253,16 @@ ## **17. Direct3D 10 Interoperability** -| **CUDA** | **HIP** | -|-----------------------------------------------------------|-------------------------------| +| **CUDA** | **HIP** |**CUDA version\***| +|-----------------------------------------------------------|-------------------------------|:----------------:| | `cudaD3D10GetDevice` | | | `cudaD3D10GetDevices` | | | `cudaGraphicsD3D10RegisterResource` | | ## **18. Direct3D 10 Interoperability [DEPRECATED]** -| **CUDA** | **HIP** | -|-----------------------------------------------------------|-------------------------------| +| **CUDA** | **HIP** |**CUDA version\***| +|-----------------------------------------------------------|-------------------------------|:----------------:| | `cudaD3D10GetDirect3DDevice` | | | `cudaD3D10MapResources` | | | `cudaD3D10RegisterResource` | | @@ -278,23 +278,23 @@ ## **19. Direct3D 11 Interoperability** -| **CUDA** | **HIP** | -|-----------------------------------------------------------|-------------------------------| +| **CUDA** | **HIP** |**CUDA version\***| +|-----------------------------------------------------------|-------------------------------|:----------------:| | `cudaD3D11GetDevice` | | | `cudaD3D11GetDevices` | | | `cudaGraphicsD3D11RegisterResource` | | ## **20. Direct3D 11 Interoperability [DEPRECATED]** -| **CUDA** | **HIP** | -|-----------------------------------------------------------|-------------------------------| +| **CUDA** | **HIP** |**CUDA version\***| +|-----------------------------------------------------------|-------------------------------|:----------------:| | `cudaD3D11GetDirect3DDevice` | | | `cudaD3D11SetDirect3DDevice` | | ## **21. VDPAU Interoperability** -| **CUDA** | **HIP** | -|-----------------------------------------------------------|-------------------------------| +| **CUDA** | **HIP** |**CUDA version\***| +|-----------------------------------------------------------|-------------------------------|:----------------:| | `cudaGraphicsVDPAURegisterOutputSurface` | | | `cudaGraphicsVDPAURegisterVideoSurface` | | | `cudaVDPAUGetDevice` | | @@ -302,25 +302,25 @@ ## **22. EGL Interoperability** -| **CUDA** | **HIP** | -|-----------------------------------------------------------|-------------------------------| -| `cudaEGLStreamConsumerAcquireFrame` | | -| `cudaEGLStreamConsumerConnect` | | -| `cudaEGLStreamConsumerConnectWithFlags` | | -| `cudaEGLStreamConsumerDisconnect` | | -| `cudaEGLStreamConsumerReleaseFrame` | | -| `cudaEGLStreamProducerConnect` | | -| `cudaEGLStreamProducerDisconnect` | | -| `cudaEGLStreamProducerPresentFrame` | | -| `cudaEGLStreamProducerReturnFrame` | | -| `cudaEventCreateFromEGLSync` | | -| `cudaGraphicsEGLRegisterImage` | | -| `cudaGraphicsResourceGetMappedEglFrame` | | +| **CUDA** | **HIP** |**CUDA version\***| +|-----------------------------------------------------------|-------------------------------|:----------------:| +| `cudaEGLStreamConsumerAcquireFrame` | | 8.0 | +| `cudaEGLStreamConsumerConnect` | | 8.0 | +| `cudaEGLStreamConsumerConnectWithFlags` | | 8.0 | +| `cudaEGLStreamConsumerDisconnect` | | 8.0 | +| `cudaEGLStreamConsumerReleaseFrame` | | 8.0 | +| `cudaEGLStreamProducerConnect` | | 8.0 | +| `cudaEGLStreamProducerDisconnect` | | 8.0 | +| `cudaEGLStreamProducerPresentFrame` | | 8.0 | +| `cudaEGLStreamProducerReturnFrame` | | 8.0 | +| `cudaEventCreateFromEGLSync` | | 9.0 | +| `cudaGraphicsEGLRegisterImage` | | 8.0 | +| `cudaGraphicsResourceGetMappedEglFrame` | | 8.0 | ## **23. Graphics Interoperability** -| **CUDA** | **HIP** | -|-----------------------------------------------------------|-------------------------------| +| **CUDA** | **HIP** |**CUDA version\***| +|-----------------------------------------------------------|-------------------------------|:----------------:| | `cudaGraphicsMapResources` | | | `cudaGraphicsResourceGetMappedMipmappedArray` | | | `cudaGraphicsResourceGetMappedPointer` | | @@ -331,8 +331,8 @@ ## **24. Texture Reference Management** -| **CUDA** | **HIP** | -|-----------------------------------------------------------|----------------------------------| +| **CUDA** | **HIP** |**CUDA version\***| +|-----------------------------------------------------------|----------------------------------|:----------------:| | `cudaBindTexture` | `hipBindTexture` | | `cudaBindTexture2D` | `hipBindTexture2D` | | `cudaBindTextureToArray` | `hipBindTextureToArray` | @@ -345,15 +345,15 @@ ## **25. Surface Reference Management** -| **CUDA** | **HIP** | -|-----------------------------------------------------------|-------------------------------| +| **CUDA** | **HIP** |**CUDA version\***| +|-----------------------------------------------------------|-------------------------------|:----------------:| | `cudaBindSurfaceToArray` | | | `cudaGetSurfaceReference` | | ## **26. Texture Object Management** -| **CUDA** | **HIP** | -|-----------------------------------------------------------|--------------------------------------| +| **CUDA** | **HIP** |**CUDA version\***| +|-----------------------------------------------------------|--------------------------------------|:----------------:| | `cudaCreateTextureObject` |`hipCreateTextureObject` | | `cudaDestroyTextureObject` |`hipDestroyTextureObject` | | `cudaGetTextureObjectResourceDesc` |`hipGetTextureObjectResourceDesc` | @@ -362,61 +362,61 @@ ## **27. Surface Object Management** -| **CUDA** | **HIP** | -|-----------------------------------------------------------|-------------------------------| +| **CUDA** | **HIP** |**CUDA version\***| +|-----------------------------------------------------------|-------------------------------|:----------------:| | `cudaCreateSurfaceObject` | `hipCreateSurfaceObject` | | `cudaDestroySurfaceObject` | `hipDestroySurfaceObject` | | `cudaGetSurfaceObjectResourceDesc` | | ## **28. Version Management** -| **CUDA** | **HIP** | -|-----------------------------------------------------------|-------------------------------| +| **CUDA** | **HIP** |**CUDA version\***| +|-----------------------------------------------------------|-------------------------------|:----------------:| | `cudaDriverGetVersion` | `hipDriverGetVersion` | | `cudaRuntimeGetVersion` | `hipRuntimeGetVersion` | ## **29. Graph Management** -| **CUDA** | **HIP** | -|-----------------------------------------------------------|-------------------------------| -| `cudaGraphAddChildGraphNode` | | -| `cudaGraphAddDependencies` | | -| `cudaGraphAddEmptyNode` | | -| `cudaGraphAddHostNode` | | -| `cudaGraphAddKernelNode` | | -| `cudaGraphAddMemcpyNode` | | -| `cudaGraphAddMemsetNode` | | -| `cudaGraphChildGraphNodeGetGraph` | | -| `cudaGraphClone` | | -| `cudaGraphCreate` | | -| `cudaGraphDestroy` | | -| `cudaGraphDestroyNode` | | -| `cudaGraphExecDestroy` | | -| `cudaGraphGetEdges` | | -| `cudaGraphGetNodes` | | -| `cudaGraphGetRootNodes` | | -| `cudaGraphHostNodeGetParams` | | -| `cudaGraphHostNodeSetParams` | | -| `cudaGraphInstantiate` | | -| `cudaGraphExecKernelNodeSetParams` | | -| `cudaGraphKernelNodeGetParams` | | -| `cudaGraphKernelNodeSetParams` | | -| `cudaGraphLaunch` | | -| `cudaGraphMemcpyNodeGetParams` | | -| `cudaGraphMemcpyNodeSetParams` | | -| `cudaGraphMemsetNodeGetParams` | | -| `cudaGraphMemsetNodeSetParams` | | -| `cudaGraphNodeFindInClone` | | -| `cudaGraphNodeGetDependencies` | | -| `cudaGraphNodeGetDependentNodes` | | -| `cudaGraphNodeGetType` | | -| `cudaGraphRemoveDependencies` | | +| **CUDA** | **HIP** |**CUDA version\***| +|-----------------------------------------------------------|-------------------------------|:----------------:| +| `cudaGraphAddChildGraphNode` | | 10.0 | +| `cudaGraphAddDependencies` | | 10.0 | +| `cudaGraphAddEmptyNode` | | 10.0 | +| `cudaGraphAddHostNode` | | 10.0 | +| `cudaGraphAddKernelNode` | | 10.0 | +| `cudaGraphAddMemcpyNode` | | 10.0 | +| `cudaGraphAddMemsetNode` | | 10.0 | +| `cudaGraphChildGraphNodeGetGraph` | | 10.0 | +| `cudaGraphClone` | | 10.0 | +| `cudaGraphCreate` | | 10.0 | +| `cudaGraphDestroy` | | 10.0 | +| `cudaGraphDestroyNode` | | 10.0 | +| `cudaGraphExecDestroy` | | 10.0 | +| `cudaGraphGetEdges` | | 10.0 | +| `cudaGraphGetNodes` | | 10.0 | +| `cudaGraphGetRootNodes` | | 10.0 | +| `cudaGraphHostNodeGetParams` | | 10.0 | +| `cudaGraphHostNodeSetParams` | | 10.0 | +| `cudaGraphInstantiate` | | 10.0 | +| `cudaGraphExecKernelNodeSetParams` | | 10.1 | +| `cudaGraphKernelNodeGetParams` | | 10.0 | +| `cudaGraphKernelNodeSetParams` | | 10.0 | +| `cudaGraphLaunch` | | 10.0 | +| `cudaGraphMemcpyNodeGetParams` | | 10.0 | +| `cudaGraphMemcpyNodeSetParams` | | 10.0 | +| `cudaGraphMemsetNodeGetParams` | | 10.0 | +| `cudaGraphMemsetNodeSetParams` | | 10.0 | +| `cudaGraphNodeFindInClone` | | 10.0 | +| `cudaGraphNodeGetDependencies` | | 10.0 | +| `cudaGraphNodeGetDependentNodes` | | 10.0 | +| `cudaGraphNodeGetType` | | 10.0 | +| `cudaGraphRemoveDependencies` | | 10.0 | ## **30. C++ API Routines** *(7.0 contains, 7.5 doesn’t)* -| **CUDA** | **HIP** | -|-----------------------------------------------------------|------------------------------------------------| +| **CUDA** | **HIP** |**CUDA version\***| +|-----------------------------------------------------------|------------------------------------------------|:----------------:| | `cudaBindSurfaceToArray` | | | `cudaBindTexture` | `hipBindTexture` | | `cudaBindTexture2D` | | @@ -451,8 +451,8 @@ ## **32. Profiler Control** -| **CUDA** | **HIP** | -|-----------------------------------------------------------|-------------------------------| +| **CUDA** | **HIP** |**CUDA version\***| +|-----------------------------------------------------------|-------------------------------|:----------------:| | `cudaProfilerInitialize` | | | `cudaProfilerStart` | `hipProfilerStart` | | `cudaProfilerStop` | `hipProfilerStop` | @@ -461,627 +461,630 @@ ## **33. Data types** -| **type** | **CUDA** | **HIP** |**HIP value** (if differs) | -|-------------:|-----------------------------------------------------|------------------------------------------------------------|---------------------------| -| struct |`cudaChannelFormatDesc` |`hipChannelFormatDesc` | -| struct |`cudaDeviceProp` |`hipDeviceProp_t` | -| struct |`cudaEglFrame` | | -| typedef |`cudaEglFrame_st` | | -| struct |`cudaEglPlaneDesc` | | -| typedef |`cudaEglPlaneDesc_st` | | -| struct |`cudaExtent` |`hipExtent` | -| struct |`cudaFuncAttributes` |`hipFuncAttributes` | -| struct |`cudaIpcEventHandle_t` |`hipIpcEventHandle_t` | -| struct |`cudaIpcMemHandle_t` |`hipIpcMemHandle_t` | -| struct |`cudaMemcpy3DParms` |`hipMemcpy3DParms` | -| struct |`cudaMemcpy3DPeerParms` | | -| struct |`cudaPitchedPtr` |`hipPitchedPtr` | -| struct |`cudaPointerAttributes` |`hipPointerAttribute_t` | -| struct |`cudaPos` |`hipPos` | -| struct |`cudaResourceDesc` |`hipResourceDesc` | -| struct |`cudaResourceViewDesc` |`hipResourceViewDesc` | -| struct |`cudaTextureDesc` |`hipTextureDesc` | -| struct |`textureReference` |`textureReference` | -| struct |`surfaceReference` | | -| struct |`CUuuid_st` | | -| enum |***`cudaCGScope`*** | | -| 0 |*`cudaCGScopeInvalid`* | | -| 1 |*`cudaCGScopeGrid`* | | -| 2 |*`cudaCGScopeMultiGrid`* | | -| enum |***`cudaChannelFormatKind`*** |***`hipChannelFormatKind`*** | -| 0 |*`cudaChannelFormatKindSigned`* |*`hipChannelFormatKindSigned`* | -| 1 |*`cudaChannelFormatKindUnsigned`* |*`hipChannelFormatKindUnsigned`* | -| 2 |*`cudaChannelFormatKindFloat`* |*`hipChannelFormatKindFloat`* | -| 3 |*`cudaChannelFormatKindNone`* |*`hipChannelFormatKindNone`* | -| enum |***`cudaComputeMode`*** |***`hipComputeMode`*** | -| 0 |*`cudaComputeModeDefault`* |*`hipComputeModeDefault`* | -| 1 |*`cudaComputeModeExclusive`* |*`hipComputeModeExclusive`* | -| 2 |*`cudaComputeModeProhibited`* |*`hipComputeModeProhibited`* | -| 3 |*`cudaComputeModeExclusiveProcess`* |*`hipComputeModeExclusiveProcess`* | -| enum |***`cudaDeviceAttr`*** |***`hipDeviceAttribute_t`*** | -| 1 |*`cudaDevAttrMaxThreadsPerBlock`* |*`hipDeviceAttributeMaxThreadsPerBlock`* | -| 2 |*`cudaDevAttrMaxBlockDimX`* |*`hipDeviceAttributeMaxBlockDimX`* | -| 3 |*`cudaDevAttrMaxBlockDimY`* |*`hipDeviceAttributeMaxBlockDimY`* | -| 4 |*`cudaDevAttrMaxBlockDimZ`* |*`hipDeviceAttributeMaxBlockDimZ`* | -| 5 |*`cudaDevAttrMaxGridDimX`* |*`hipDeviceAttributeMaxGridDimX`* | -| 6 |*`cudaDevAttrMaxGridDimY`* |*`hipDeviceAttributeMaxGridDimY`* | -| 7 |*`cudaDevAttrMaxGridDimZ`* |*`hipDeviceAttributeMaxGridDimZ`* | -| 8 |*`cudaDevAttrMaxSharedMemoryPerBlock`* |*`hipDeviceAttributeMaxSharedMemoryPerBlock`* | -| 9 |*`cudaDevAttrTotalConstantMemory`* |*`hipDeviceAttributeTotalConstantMemory`* | -| 10 |*`cudaDevAttrWarpSize`* |*`hipDeviceAttributeWarpSize`* | -| 11 |*`cudaDevAttrMaxPitch`* | | -| 12 |*`cudaDevAttrMaxRegistersPerBlock`* |*`hipDeviceAttributeMaxRegistersPerBlock`* | -| 13 |*`cudaDevAttrClockRate`* |*`hipDeviceAttributeClockRate`* | -| 14 |*`cudaDevAttrTextureAlignment`* | | -| 15 |*`cudaDevAttrGpuOverlap`* | | -| 16 |*`cudaDevAttrMultiProcessorCount`* |*`hipDeviceAttributeMultiprocessorCount`* | -| 17 |*`cudaDevAttrKernelExecTimeout`* | | -| 18 |*`cudaDevAttrIntegrated`* |*`hipDeviceAttributeIntegrated`* | -| 19 |*`cudaDevAttrCanMapHostMemory`* | | -| 20 |*`cudaDevAttrComputeMode`* |*`hipDeviceAttributeComputeMode`* | -| 21 |*`cudaDevAttrMaxTexture1DWidth`* |*`hipDeviceAttributeMaxTexture1DWidth`* | -| 22 |*`cudaDevAttrMaxTexture2DWidth`* |*`hipDeviceAttributeMaxTexture2DWidth`* | -| 23 |*`cudaDevAttrMaxTexture2DHeight`* |*`hipDeviceAttributeMaxTexture2DHeight`* | -| 24 |*`cudaDevAttrMaxTexture3DWidth`* |*`hipDeviceAttributeMaxTexture3DWidth`* | -| 25 |*`cudaDevAttrMaxTexture3DHeight`* |*`hipDeviceAttributeMaxTexture3DHeight`* | -| 26 |*`cudaDevAttrMaxTexture3DDepth`* |*`hipDeviceAttributeMaxTexture3DDepth`* | -| 27 |*`cudaDevAttrMaxTexture2DLayeredWidth`* | | -| 28 |*`cudaDevAttrMaxTexture2DLayeredHeight`* | | -| 29 |*`cudaDevAttrMaxTexture2DLayeredLayers`* | | -| 30 |*`cudaDevAttrSurfaceAlignment`* | | -| 31 |*`cudaDevAttrConcurrentKernels`* |*`hipDeviceAttributeConcurrentKernels`* | -| 32 |*`cudaDevAttrEccEnabled`* | | -| 33 |*`cudaDevAttrPciBusId`* |*`hipDeviceAttributePciBusId`* | -| 34 |*`cudaDevAttrPciDeviceId`* |*`hipDeviceAttributePciDeviceId`* | -| 35 |*`cudaDevAttrTccDriver`* | | -| 36 |*`cudaDevAttrMemoryClockRate`* |*`hipDeviceAttributeMemoryClockRate`* | -| 37 |*`cudaDevAttrGlobalMemoryBusWidth`* |*`hipDeviceAttributeMemoryBusWidth`* | -| 38 |*`cudaDevAttrL2CacheSize`* |*`hipDeviceAttributeL2CacheSize`* | -| 39 |*`cudaDevAttrMaxThreadsPerMultiProcessor`* |*`hipDeviceAttributeMaxThreadsPerMultiProcessor`* | -| 40 |*`cudaDevAttrAsyncEngineCount`* | | -| 41 |*`cudaDevAttrUnifiedAddressing`* | | -| 42 |*`cudaDevAttrMaxTexture1DLayeredWidth`* | | -| 43 |*`cudaDevAttrMaxTexture1DLayeredLayers`* | | -| 44 | | | -| 45 |*`cudaDevAttrMaxTexture2DGatherWidth`* | | -| 46 |*`cudaDevAttrMaxTexture2DGatherHeight`* | | -| 47 |*`cudaDevAttrMaxTexture3DWidthAlt`* | | -| 48 |*`cudaDevAttrMaxTexture3DHeightAlt`* | | -| 49 |*`cudaDevAttrMaxTexture3DDepthAlt`* | | -| 50 |*`cudaDevAttrPciDomainId`* | | -| 51 |*`cudaDevAttrTexturePitchAlignment`* | | -| 52 |*`cudaDevAttrMaxTextureCubemapWidth`* | | -| 53 |*`cudaDevAttrMaxTextureCubemapLayeredWidth`* | | -| 54 |*`cudaDevAttrMaxTextureCubemapLayeredLayers`* | | -| 55 |*`cudaDevAttrMaxSurface1DWidth`* | | -| 56 |*`cudaDevAttrMaxSurface2DWidth`* | | -| 57 |*`cudaDevAttrMaxSurface2DHeight`* | | -| 58 |*`cudaDevAttrMaxSurface3DWidth`* | | -| 59 |*`cudaDevAttrMaxSurface3DHeight`* | | -| 60 |*`cudaDevAttrMaxSurface3DDepth`* | | -| 61 |*`cudaDevAttrMaxSurface1DLayeredWidth`* | | -| 62 |*`cudaDevAttrMaxSurface1DLayeredLayers`* | | -| 63 |*`cudaDevAttrMaxSurface2DLayeredWidth`* | | -| 64 |*`cudaDevAttrMaxSurface2DLayeredHeight`* | | -| 65 |*`cudaDevAttrMaxSurface2DLayeredLayers`* | | -| 66 |*`cudaDevAttrMaxSurfaceCubemapWidth`* | | -| 67 |*`cudaDevAttrMaxSurfaceCubemapLayeredWidth`* | | -| 68 |*`cudaDevAttrMaxSurfaceCubemapLayeredLayers`* | | -| 69 |*`cudaDevAttrMaxTexture1DLinearWidth`* | | -| 70 |*`cudaDevAttrMaxTexture2DLinearWidth`* | | -| 71 |*`cudaDevAttrMaxTexture2DLinearHeight`* | | -| 72 |*`cudaDevAttrMaxTexture2DLinearPitch`* | | -| 73 |*`cudaDevAttrMaxTexture2DMipmappedWidth* | | -| 74 |*`cudaDevAttrMaxTexture2DMipmappedHeight`* | | -| 75 |*`cudaDevAttrComputeCapabilityMajor`* |*`hipDeviceAttributeComputeCapabilityMajor`* | -| 76 |*`cudaDevAttrComputeCapabilityMinor`* |*`hipDeviceAttributeComputeCapabilityMinor`* | -| 77 |*`cudaDevAttrMaxTexture1DMipmappedWidth`* | | -| 78 |*`cudaDevAttrStreamPrioritiesSupported`* | | -| 79 |*`cudaDevAttrGlobalL1CacheSupported`* | | -| 80 |*`cudaDevAttrLocalL1CacheSupported`* | | -| 81 |*`cudaDevAttrMaxSharedMemoryPerMultiprocessor`* |*`hipDeviceAttributeMaxSharedMemoryPerMultiprocessor`* | -| 82 |*`cudaDevAttrMaxRegistersPerMultiprocessor`* | | -| 83 |*`cudaDevAttrManagedMemory`* | | -| 84 |*`cudaDevAttrIsMultiGpuBoard`* |*`hipDeviceAttributeIsMultiGpuBoard`* | -| 85 |*`cudaDevAttrMultiGpuBoardGroupID`* | | -| 86 |*`cudaDevAttrHostNativeAtomicSupported`* | | -| 87 |*`cudaDevAttrSingleToDoublePrecisionPerfRatio`* | | -| 88 |*`cudaDevAttrPageableMemoryAccess`* | | -| 89 |*`cudaDevAttrConcurrentManagedAccess`* | | -| 90 |*`cudaDevAttrComputePreemptionSupported`* | | -| 91 |*`cudaDevAttrCanUseHostPointerForRegisteredMem`* | | -| 92 |*`cudaDevAttrReserved92`* | | -| 93 |*`cudaDevAttrReserved93`* | | -| 94 |*`cudaDevAttrReserved94`* | | -| 95 |*`cudaDevAttrCooperativeLaunch`* | | -| 96 |*`cudaDevAttrCooperativeMultiDeviceLaunch`* | | -| 97 |*`cudaDevAttrMaxSharedMemoryPerBlockOptin`* | | -| 98 |*`cudaDevAttrCanFlushRemoteWrites`* | | -| 99 |*`cudaDevAttrHostRegisterSupported`* | | -| 100 |*`cudaDevAttrPageableMemoryAccessUsesHostPageTables`*| | -| 101 |*`cudaDevAttrDirectManagedMemAccessFromHost`* | | -| enum |***`cudaDeviceP2PAttr`*** | | -| 1 |*`cudaDevP2PAttrPerformanceRank`* | | -| 2 |*`cudaDevP2PAttrAccessSupported`* | | -| 3 |*`cudaDevP2PAttrNativeAtomicSupported`* | | -| 4 |*`cudaDevP2PAttrCudaArrayAccessSupported`* | | -| enum |***`cudaEglColorFormat`*** | | -| 0 |*`cudaEglColorFormatYUV420Planar`* | | -| 1 |*`cudaEglColorFormatYUV420SemiPlanar`* | | -| 2 |*`cudaEglColorFormatYUV422Planar`* | | -| 3 |*`cudaEglColorFormatYUV422SemiPlanar`* | | -| 4 |*`cudaEglColorFormatRGB`* | | -| 5 |*`cudaEglColorFormatBGR`* | | -| 6 |*`cudaEglColorFormatARGB`* | | -| 7 |*`cudaEglColorFormatRGBA`* | | -| 8 |*`cudaEglColorFormatL`* | | -| 9 |*`cudaEglColorFormatR`* | | -| 10 |*`cudaEglColorFormatYUV444Planar`* | | -| 11 |*`cudaEglColorFormatYUV444SemiPlanar`* | | -| 12 |*`cudaEglColorFormatYUYV422`* | | -| 13 |*`cudaEglColorFormatUYVY422`* | | -| 14 |*`cudaEglColorFormatABGR`* | | -| 15 |*`cudaEglColorFormatBGRA`* | | -| 16 |*`cudaEglColorFormatA`* | | -| 17 |*`cudaEglColorFormatRG`* | | -| 18 |*`cudaEglColorFormatAYUV`* | | -| 19 |*`cudaEglColorFormatYVU444SemiPlanar`* | | -| 20 |*`cudaEglColorFormatYVU422SemiPlanar`* | | -| 21 |*`cudaEglColorFormatYVU420SemiPlanar`* | | -| 22 |*`cudaEglColorFormatY10V10U10_444SemiPlanar`* | | -| 23 |*`cudaEglColorFormatY10V10U10_420SemiPlanar`* | | -| 24 |*`cudaEglColorFormatY12V12U12_444SemiPlanar`* | | -| 25 |*`cudaEglColorFormatY12V12U12_420SemiPlanar`* | | -| 26 |*`cudaEglColorFormatVYUY_ER`* | | -| 27 |*`cudaEglColorFormatUYVY_ER`* | | -| 28 |*`cudaEglColorFormatYUYV_ER`* | | -| 29 |*`cudaEglColorFormatYVYU_ER`* | | -| 30 |*`cudaEglColorFormatYUV_ER`* | | -| 31 |*`cudaEglColorFormatYUVA_ER`* | | -| 32 |*`cudaEglColorFormatAYUV_ER`* | | -| 33 |*`cudaEglColorFormatYUV444Planar_ER`* | | -| 34 |*`cudaEglColorFormatYUV422Planar_ER`* | | -| 35 |*`cudaEglColorFormatYUV420Planar_ER`* | | -| 36 |*`cudaEglColorFormatYUV444SemiPlanar_ER`* | | -| 37 |*`cudaEglColorFormatYUV422SemiPlanar_ER`* | | -| 38 |*`cudaEglColorFormatYUV420SemiPlanar_ER`* | | -| 39 |*`cudaEglColorFormatYVU444Planar_ER`* | | -| 40 |*`cudaEglColorFormatYVU422Planar_ER`* | | -| 41 |*`cudaEglColorFormatYVU420Planar_ER`* | | -| 42 |*`cudaEglColorFormatYVU444SemiPlanar_ER`* | | -| 43 |*`cudaEglColorFormatYVU422SemiPlanar_ER`* | | -| 44 |*`cudaEglColorFormatYVU420SemiPlanar_ER`* | | -| 45 |*`cudaEglColorFormatBayerRGGB`* | | -| 46 |*`cudaEglColorFormatBayerBGGR`* | | -| 47 |*`cudaEglColorFormatBayerGRBG`* | | -| 48 |*`cudaEglColorFormatBayerGBRG`* | | -| 49 |*`cudaEglColorFormatBayer10RGGB`* | | -| 50 |*`cudaEglColorFormatBayer10BGGR`* | | -| 51 |*`cudaEglColorFormatBayer10GRBG`* | | -| 52 |*`cudaEglColorFormatBayer10GBRG`* | | -| 53 |*`cudaEglColorFormatBayer12RGGB`* | | -| 54 |*`cudaEglColorFormatBayer12BGGR`* | | -| 55 |*`cudaEglColorFormatBayer12GRBG`* | | -| 56 |*`cudaEglColorFormatBayer12GBRG`* | | -| 57 |*`cudaEglColorFormatBayer14RGGB`* | | -| 58 |*`cudaEglColorFormatBayer14BGGR`* | | -| 59 |*`cudaEglColorFormatBayer14GRBG`* | | -| 60 |*`cudaEglColorFormatBayer14GBRG`* | | -| 61 |*`cudaEglColorFormatBayer20RGGB`* | | -| 62 |*`cudaEglColorFormatBayer20BGGR`* | | -| 63 |*`cudaEglColorFormatBayer20GRBG`* | | -| 64 |*`cudaEglColorFormatBayer20GBRG`* | | -| 65 |*`cudaEglColorFormatYVU444Planar`* | | -| 66 |*`cudaEglColorFormatYVU422Planar`* | | -| 67 |*`cudaEglColorFormatYVU420Planar`* | | -| 68 |*`cudaEglColorFormatBayerIspRGGB`* | | -| 69 |*`cudaEglColorFormatBayerIspBGGR`* | | -| 70 |*`cudaEglColorFormatBayerIspGRBG`* | | -| 71 |*`cudaEglColorFormatBayerIspGBRG`* | | -| enum |***`cudaEglFrameType`*** | | -| 0 |*`cudaEglFrameTypeArray`* | | -| 1 |*`cudaEglFrameTypePitch`* | | -| enum |***`cudaExternalMemoryHandleType`*** | | -| 1 |*`cudaExternalMemoryHandleTypeOpaqueFd`* | | -| 2 |*`cudaExternalMemoryHandleTypeOpaqueWin32`* | | -| 3 |*`cudaExternalMemoryHandleTypeOpaqueWin32Kmt`* | | -| 4 |*`cudaExternalMemoryHandleTypeD3D12Heap`* | | -| 5 |*`cudaExternalMemoryHandleTypeD3D12Resource`* | | -| enum |***`cudaExternalSemaphoreHandleType`*** | | -| 1 |*`cudaExternalSemaphoreHandleTypeOpaqueFd`* | | -| 2 |*`cudaExternalSemaphoreHandleTypeOpaqueWin32`* | | -| 3 |*`cudaExternalSemaphoreHandleTypeOpaqueWin32Kmt`* | | -| 4 |*`cudaExternalSemaphoreHandleTypeD3D12Fence`* | | -| enum |***`cudaFuncAttribute`*** | | -| 8 |*`cudaFuncAttributeMaxDynamicSharedMemorySize`* | | -| 9 |*`cudaFuncAttributePreferredSharedMemoryCarveout`* | | -| 10 |*`cudaFuncAttributeMax`* | | -| enum |***`cudaEglResourceLocationFlags`*** | | -| 0x00 |*`cudaEglResourceLocationSysmem`* | | -| 0x01 |*`cudaEglResourceLocationVidmem`* | | -| enum |***`cudaError`*** |***`hipError_t`*** | -| typedef |***`cudaError_t`*** |***`hipError_t`*** | -| 0 |*`cudaSuccess`* |*`hipSuccess`* | -| 1 |*`cudaErrorMissingConfiguration`* |*`hipErrorMissingConfiguration`* | 1001 | -| 2 |*`cudaErrorMemoryAllocation`* |*`hipErrorMemoryAllocation`* | 1002 | -| 3 |*`cudaErrorInitializationError`* |*`hipErrorInitializationError`* | 1003 | -| 4 |*`cudaErrorLaunchFailure`* |*`hipErrorLaunchFailure`* | 1004 | -| 5 |*`cudaErrorPriorLaunchFailure`* |*`hipErrorPriorLaunchFailure`* | 1005 | -| 6 |*`cudaErrorLaunchTimeout`* |*`hipErrorLaunchTimeOut`* | 1006 | -| 7 |*`cudaErrorLaunchOutOfResources`* |*`hipErrorLaunchOutOfResources`* | 1007 | -| 8 |*`cudaErrorInvalidDeviceFunction`* |*`hipErrorInvalidDeviceFunction`* | 1008 | -| 9 |*`cudaErrorInvalidConfiguration`* |*`hipErrorInvalidConfiguration`* | 1009 | -| 10 |*`cudaErrorInvalidDevice`* |*`hipErrorInvalidDevice`* | 1010 | -| 11 |*`cudaErrorInvalidValue`* |*`hipErrorInvalidValue`* | 1011 | -| 12 |*`cudaErrorInvalidPitchValue`* | | -| 13 |*`cudaErrorInvalidSymbol`* |*`hipErrorInvalidSymbol`* | 701 | -| 14 |*`cudaErrorMapBufferObjectFailed`* |*`hipErrorMapFailed`* | 205 | -| 15 |*`cudaErrorUnmapBufferObjectFailed`* |*`hipErrorUnmapFailed`* | 206 | -| 16 |*`cudaErrorInvalidHostPointer`* | | -| 17 |*`cudaErrorInvalidDevicePointer`* |*`hipErrorInvalidDevicePointer`* | 1017 | -| 18 |*`cudaErrorInvalidTexture`* | | -| 19 |*`cudaErrorInvalidTextureBinding`* | | -| 20 |*`cudaErrorInvalidChannelDescriptor`* | | -| 21 |*`cudaErrorInvalidMemcpyDirection`* |*`hipErrorInvalidMemcpyDirection`* | 1021 | -| 22 |*`cudaErrorAddressOfConstant`* | | -| 23 |*`cudaErrorTextureFetchFailed`* | | -| 24 |*`cudaErrorTextureNotBound`* | | -| 25 |*`cudaErrorSynchronizationError`* | | -| 26 |*`cudaErrorInvalidFilterSetting`* | | -| 27 |*`cudaErrorInvalidNormSetting`* | | -| 28 |*`cudaErrorMixedDeviceExecution`* | | -| 29 |*`cudaErrorCudartUnloading`* |*`hipErrorDeinitialized`* | 4 | -| 30 |*`cudaErrorUnknown`* |*`hipErrorUnknown`* | 1030 | -| 31 |*`cudaErrorNotYetImplemented`* | | -| 32 |*`cudaErrorMemoryValueTooLarge`* | | -| 33 |*`cudaErrorInvalidResourceHandle`* |*`hipErrorInvalidResourceHandle`* | 1033 | -| 34 |*`cudaErrorNotReady`* |*`hipErrorNotReady`* | 1034 | -| 35 |*`cudaErrorInsufficientDriver`* |*`hipErrorInsufficientDriver`* | -| 36 |*`cudaErrorSetOnActiveProcess`* | | -| 37 |*`cudaErrorInvalidSurface`* | | -| 38 |*`cudaErrorNoDevice`* |*`hipErrorNoDevice`* | 1038 | -| 39 |*`cudaErrorECCUncorrectable`* |*`hipErrorECCNotCorrectable`* | 214 | -| 40 |*`cudaErrorSharedObjectSymbolNotFound`* |*`hipErrorSharedObjectSymbolNotFound`* | 302 | -| 41 |*`cudaErrorSharedObjectInitFailed`* |*`hipErrorSharedObjectInitFailed`* | 303 | -| 42 |*`cudaErrorUnsupportedLimit`* |*`hipErrorUnsupportedLimit`* | 215 | -| 43 |*`cudaErrorDuplicateVariableName`* | | -| 44 |*`cudaErrorDuplicateTextureName`* | | -| 45 |*`cudaErrorDuplicateSurfaceName`* | | -| 46 |*`cudaErrorDevicesUnavailable`* | | -| 47 |*`cudaErrorInvalidKernelImage`* |*`hipErrorInvalidImage`* | 200 | -| 48 |*`cudaErrorNoKernelImageForDevice`* |*`hipErrorNoBinaryForGpu`* | 209 | -| 49 |*`cudaErrorIncompatibleDriverContext`* | | -| 50 |*`cudaErrorPeerAccessAlreadyEnabled`* |*`hipErrorPeerAccessAlreadyEnabled`* | 1050 | -| 51 |*`cudaErrorPeerAccessNotEnabled`* |*`hipErrorPeerAccessNotEnabled`* | 1051 | -| 52 | | | -| 53 | | | -| 54 |*`cudaErrorDeviceAlreadyInUse`* | | -| 55 |*`cudaErrorProfilerDisabled`* |*`hipErrorProfilerDisabled`* | 5 | -| 56 |*`cudaErrorProfilerNotInitialized`* |*`hipErrorProfilerNotInitialized`* | 6 | -| 57 |*`cudaErrorProfilerAlreadyStarted`* |*`hipErrorProfilerAlreadyStarted`* | 7 | -| 58 |*`cudaErrorProfilerAlreadyStopped`* |*`hipErrorProfilerAlreadyStopped`* | 8 | -| 59 |*`cudaErrorAssert`* |*`hipErrorAssert`* | 1081 | -| 60 |*`cudaErrorTooManyPeers`* | | -| 61 |*`cudaErrorHostMemoryAlreadyRegistered`* |*`hipErrorHostMemoryAlreadyRegistered`* | 1061 | -| 62 |*`cudaErrorHostMemoryNotRegistered`* |*`hipErrorHostMemoryNotRegistered`* | 1062 | -| 63 |*`cudaErrorOperatingSystem`* |*`hipErrorOperatingSystem`* | 304 | -| 64 |*`cudaErrorPeerAccessUnsupported`* |*`hipErrorPeerAccessUnsupported`* | 217 | -| 65 |*`cudaErrorLaunchMaxDepthExceeded`* | | -| 66 |*`cudaErrorLaunchFileScopedTex`* | | -| 67 |*`cudaErrorLaunchFileScopedSurf`* | | -| 68 |*`cudaErrorSyncDepthExceeded`* | | -| 69 |*`cudaErrorLaunchPendingCountExceeded`* | | -| 70 |*`cudaErrorNotPermitted`* | | -| 71 |*`cudaErrorNotSupported`* | | -| 72 |*`cudaErrorHardwareStackError`* | | -| 73 |*`cudaErrorIllegalInstruction`* | | -| 74 |*`cudaErrorMisalignedAddress`* | | -| 75 |*`cudaErrorInvalidAddressSpace`* | | -| 76 |*`cudaErrorInvalidPc`* | | -| 77 |*`cudaErrorIllegalAddress`* |*`hipErrorIllegalAddress`* | 700 | -| 78 |*`cudaErrorInvalidPtx`* |*`hipErrorInvalidKernelFile`* | 218 | -| 79 |*`cudaErrorInvalidGraphicsContext`* |*`hipErrorInvalidGraphicsContext`* | 219 | -| 80 |*`cudaErrorNvlinkUncorrectable`* | | -| 81 |*`cudaErrorJitCompilerNotFound`* | | -| 82 |*`cudaErrorCooperativeLaunchTooLarge`* | | -| 83 |*`cudaErrorSystemNotReady`* | | -| 84 |*`cudaErrorIllegalState`* | | -| 127 |*`cudaErrorStartupFailure`* | | -| 803 |*`cudaErrorSystemDriverMismatch`* | | -| 804 |*`cudaErrorCompatNotSupportedOnDevice`* | | -| 900 |*`cudaErrorStreamCaptureUnsupported`* | | -| 901 |*`cudaErrorStreamCaptureInvalidated`* | | -| 902 |*`cudaErrorStreamCaptureMerge`* | | -| 903 |*`cudaErrorStreamCaptureUnmatched`* | | -| 904 |*`cudaErrorStreamCaptureUnjoined`* | | -| 905 |*`cudaErrorStreamCaptureIsolation`* | | -| 906 |*`cudaErrorStreamCaptureImplicit`* | | -| 907 |*`cudaErrorCapturedEvent`* | | -| 908 |*`cudaErrorStreamCaptureWrongThread`* | | -| 10000 |*`cudaErrorApiFailureBase`* | | -| enum |***`cudaFuncCache`*** |***`hipFuncCache_t`*** | -| 0 |*`cudaFuncCachePreferNone`* |*`hipFuncCachePreferNone`* | -| 1 |*`cudaFuncCachePreferShared`* |*`hipFuncCachePreferShared`* | -| 2 |*`cudaFuncCachePreferL1`* |*`hipFuncCachePreferL1`* | -| 3 |*`cudaFuncCachePreferEqual`* |*`hipFuncCachePreferEqual`* | -| enum |***`cudaGraphicsCubeFace`*** | | -| 0x00 |*`cudaGraphicsCubeFacePositiveX`* | | -| 0x01 |*`cudaGraphicsCubeFaceNegativeX`* | | -| 0x02 |*`cudaGraphicsCubeFacePositiveY`* | | -| 0x03 |*`cudaGraphicsCubeFaceNegativeY`* | | -| 0x04 |*`cudaGraphicsCubeFacePositiveZ`* | | -| 0x05 |*`cudaGraphicsCubeFaceNegativeZ`* | | -| enum |***`cudaGraphicsMapFlags`*** | | -| 0 |*`cudaGraphicsMapFlagsNone`* | | -| 1 |*`cudaGraphicsMapFlagsReadOnly`* | | -| 2 |*`cudaGraphicsMapFlagsWriteDiscard`* | | -| enum |***`cudaGraphicsRegisterFlags`*** | | -| 0 |*`cudaGraphicsRegisterFlagsNone`* | | -| 1 |*`cudaGraphicsRegisterFlagsReadOnly`* | | -| 2 |*`cudaGraphicsRegisterFlagsWriteDiscard`* | | -| 4 |*`cudaGraphicsRegisterFlagsSurfaceLoadStore`* | | -| 8 |*`cudaGraphicsRegisterFlagsTextureGather`* | | -| enum |***`cudaGraphNodeType`*** | | -| 0x00 |*`cudaGraphNodeTypeKernel`* | | -| 0x01 |*`cudaGraphNodeTypeMemcpy`* | | -| 0x02 |*`cudaGraphNodeTypeMemset`* | | -| 0x03 |*`cudaGraphNodeTypeHost`* | | -| 0x04 |*`cudaGraphNodeTypeGraph`* | | -| 0x05 |*`cudaGraphNodeTypeEmpty`* | | -| |*`cudaGraphNodeTypeCount`* | | -| enum |***`cudaLimit`*** |***`hipLimit_t`*** | -| 0x00 |*`cudaLimitStackSize`* | | -| 0x01 |*`cudaLimitPrintfFifoSize`* | | -| 0x02 |*`cudaLimitMallocHeapSize`* |*`hipLimitMallocHeapSize`* | -| 0x03 |*`cudaLimitDevRuntimeSyncDepth`* | | -| 0x04 |*`cudaLimitDevRuntimePendingLaunchCount`* | | -| 0x05 |*`cudaLimitMaxL2FetchGranularity`* | | -| enum |***`cudaMemcpyKind`*** |***`hipMemcpyKind`*** | -| 0 |*`cudaMemcpyHostToHost`* |*`hipMemcpyHostToHost`* | -| 1 |*`cudaMemcpyHostToDevice`* |*`hipMemcpyHostToDevice`* | -| 2 |*`cudaMemcpyDeviceToHost`* |*`hipMemcpyDeviceToHost`* | -| 3 |*`cudaMemcpyDeviceToDevice`* |*`hipMemcpyDeviceToDevice`* | -| 4 |*`cudaMemcpyDefault`* |*`hipMemcpyDefault`* | -| enum |***`cudaMemoryAdvise`*** | | -| 1 |*`cudaMemAdviseSetReadMostly`* | | -| 2 |*`cudaMemAdviseUnsetReadMostly`* | | -| 3 |*`cudaMemAdviseSetPreferredLocation`* | | -| 4 |*`cudaMemAdviseUnsetPreferredLocation`* | | -| 5 |*`cudaMemAdviseSetAccessedBy`* | | -| 6 |*`cudaMemAdviseUnsetAccessedBy`* | | -| enum |***`cudaMemoryType`*** | | -| 0 |*`cudaMemoryTypeUnregistered`* | | -| 1 |*`cudaMemoryTypeHost`* | | -| 2 |*`cudaMemoryTypeDevice`* | | -| 3 |*`cudaMemoryTypeManaged`* | | -| enum |***`cudaMemRangeAttribute`*** | | -| 1 |*`cudaMemRangeAttributeReadMostly`* | | -| 2 |*`cudaMemRangeAttributePreferredLocation`* | | -| 3 |*`cudaMemRangeAttributeAccessedBy`* | | -| 4 |*`cudaMemRangeAttributeLastPrefetchLocation`* | | -| enum |***`cudaResourceType`*** |***`hipResourceType`*** | -| 0x00 |*`cudaResourceTypeArray`* |*`hipResourceTypeArray`* | -| 0x01 |*`cudaResourceTypeMipmappedArray`* |*`hipResourceTypeMipmappedArray`* | -| 0x02 |*`cudaResourceTypeLinear`* |*`hipResourceTypeLinear`* | -| 0x03 |*`cudaResourceTypePitch2D`* |*`hipResourceTypePitch2D`* | -| enum |***`cudaResourceViewFormat`*** |***`hipResourceViewFormat`*** | -| 0x00 |*`cudaResViewFormatNone`* |*`hipResViewFormatNone`* | -| 0x01 |*`cudaResViewFormatUnsignedChar1`* |*`hipResViewFormatUnsignedChar1`* | -| 0x02 |*`cudaResViewFormatUnsignedChar2`* |*`hipResViewFormatUnsignedChar2`* | -| 0x03 |*`cudaResViewFormatUnsignedChar4`* |*`hipResViewFormatUnsignedChar4`* | -| 0x04 |*`cudaResViewFormatSignedChar1`* |*`hipResViewFormatSignedChar1`* | -| 0x05 |*`cudaResViewFormatSignedChar2`* |*`hipResViewFormatSignedChar2`* | -| 0x06 |*`cudaResViewFormatSignedChar4`* |*`hipResViewFormatSignedChar4`* | -| 0x07 |*`cudaResViewFormatUnsignedShort1`* |*`hipResViewFormatUnsignedShort1`* | -| 0x08 |*`cudaResViewFormatUnsignedShort2`* |*`hipResViewFormatUnsignedShort2`* | -| 0x09 |*`cudaResViewFormatUnsignedShort4`* |*`hipResViewFormatUnsignedShort4`* | -| 0x0a |*`cudaResViewFormatSignedShort1`* |*`hipResViewFormatSignedShort1`* | -| 0x0b |*`cudaResViewFormatSignedShort2`* |*`hipResViewFormatSignedShort2`* | -| 0x0c |*`cudaResViewFormatSignedShort4`* |*`hipResViewFormatSignedShort4`* | -| 0x0d |*`cudaResViewFormatUnsignedInt1`* |*`hipResViewFormatUnsignedInt1`* | -| 0x0e |*`cudaResViewFormatUnsignedInt2`* |*`hipResViewFormatUnsignedInt2`* | -| 0x0f |*`cudaResViewFormatUnsignedInt4`* |*`hipResViewFormatUnsignedInt4`* | -| 0x10 |*`cudaResViewFormatSignedInt1`* |*`hipResViewFormatSignedInt1`* | -| 0x11 |*`cudaResViewFormatSignedInt2`* |*`hipResViewFormatSignedInt2`* | -| 0x12 |*`cudaResViewFormatSignedInt4`* |*`hipResViewFormatSignedInt4`* | -| 0x13 |*`cudaResViewFormatHalf1`* |*`hipResViewFormatHalf1`* | -| 0x14 |*`cudaResViewFormatHalf2`* |*`hipResViewFormatHalf2`* | -| 0x15 |*`cudaResViewFormatHalf4`* |*`hipResViewFormatHalf4`* | -| 0x16 |*`cudaResViewFormatFloat1`* |*`hipResViewFormatFloat1`* | -| 0x17 |*`cudaResViewFormatFloat2`* |*`hipResViewFormatFloat2`* | -| 0x18 |*`cudaResViewFormatFloat4`* |*`hipResViewFormatFloat4`* | -| 0x19 |*`cudaResViewFormatUnsignedBlockCompressed1`* |*`hipResViewFormatUnsignedBlockCompressed1`* | -| 0x1a |*`cudaResViewFormatUnsignedBlockCompressed2`* |*`hipResViewFormatUnsignedBlockCompressed2`* | -| 0x1b |*`cudaResViewFormatUnsignedBlockCompressed3`* |*`hipResViewFormatUnsignedBlockCompressed3`* | -| 0x1c |*`cudaResViewFormatUnsignedBlockCompressed4`* |*`hipResViewFormatUnsignedBlockCompressed4`* | -| 0x1d |*`cudaResViewFormatSignedBlockCompressed4`* |*`hipResViewFormatSignedBlockCompressed4`* | -| 0x1e |*`cudaResViewFormatUnsignedBlockCompressed5`* |*`hipResViewFormatUnsignedBlockCompressed5`* | -| 0x1f |*`cudaResViewFormatSignedBlockCompressed5`* |*`hipResViewFormatSignedBlockCompressed5`* | -| 0x20 |*`cudaResViewFormatUnsignedBlockCompressed6H`* |*`hipResViewFormatUnsignedBlockCompressed6H`* | -| 0x21 |*`cudaResViewFormatSignedBlockCompressed6H`* |*`hipResViewFormatSignedBlockCompressed6H`* | -| 0x22 |*`cudaResViewFormatUnsignedBlockCompressed7`* |*`hipResViewFormatUnsignedBlockCompressed7`* | -| enum |***`cudaSharedMemConfig`*** |***`hipSharedMemConfig`*** | -| 0 |*`cudaSharedMemBankSizeDefault`* |*`hipSharedMemBankSizeDefault`* | -| 1 |*`cudaSharedMemBankSizeFourByte`* |*`hipSharedMemBankSizeFourByte`* | -| 2 |*`cudaSharedMemBankSizeEightByte`* |*`hipSharedMemBankSizeEightByte`* | -| enum |***`cudaSharedCarveout`*** | | -| -1 |*`cudaSharedmemCarveoutDefault`* | | -| 100 |*`cudaSharedmemCarveoutMaxShared`* | | -| 0 |*`cudaSharedmemCarveoutMaxShared`* | | -| enum |***`cudaStreamCaptureStatus`*** | | -| 0 |*`cudaStreamCaptureStatusNone`* | | -| 1 |*`cudaStreamCaptureStatusActive`* | | -| 2 |*`cudaStreamCaptureStatusInvalidated`* | | -| enum |***`cudaStreamCaptureMode`*** | | -| 0 |*`cudaStreamCaptureModeGlobal`* | | -| 1 |*`cudaStreamCaptureModeThreadLocal`* | | -| 2 |*`cudaStreamCaptureModeRelaxed`* | | -| enum |***`cudaSurfaceBoundaryMode`*** |***`hipSurfaceBoundaryMode`*** | -| 0 |*`cudaBoundaryModeZero`* |*`hipBoundaryModeZero`* | -| 1 |*`cudaBoundaryModeClamp`* |*`hipBoundaryModeClamp`* | -| 2 |*`cudaBoundaryModeTrap`* |*`hipBoundaryModeTrap`* | -| enum |***`cudaSurfaceFormatMode`*** | | -| 0 |*`cudaFormatModeForced`* | | -| 1 |*`cudaFormatModeAuto`* | | -| enum |***`cudaTextureAddressMode`*** |***`hipTextureAddressMode`*** | -| 0 |*`cudaAddressModeWrap`* |*`hipAddressModeWrap`* | -| 1 |*`cudaAddressModeClamp`* |*`hipAddressModeClamp`* | -| 2 |*`cudaAddressModeMirror`* |*`hipAddressModeMirror`* | -| 3 |*`cudaAddressModeBorder`* |*`hipAddressModeBorder`* | -| enum |***`cudaTextureFilterMode`*** |***`hipTextureFilterMode`*** | -| 0 |*`cudaFilterModePoint`* |*`hipFilterModePoint`* | -| 1 |*`cudaFilterModeLinear`* |*`hipFilterModeLinear`* | -| enum |***`cudaTextureReadMode`*** |***`hipTextureReadMode`*** | -| 0 |*`cudaReadModeElementType`* |*`hipReadModeElementType`* | -| 1 |*`cudaReadModeNormalizedFloat`* |*`hipReadModeNormalizedFloat`* | -| enum |***`cudaGLDeviceList`*** | | -| 1 |*`cudaGLDeviceListAll`* | | -| 2 |*`cudaGLDeviceListCurrentFrame`* | | -| 3 |*`cudaGLDeviceListNextFrame`* | | -| enum |***`cudaGLMapFlags`*** | | -| 0 |*`cudaGLMapFlagsNone`* | | -| 1 |*`cudaGLMapFlagsReadOnly`* | | -| 2 |*`cudaGLMapFlagsWriteDiscard`* | | -| enum |***`cudaD3D9DeviceList`*** | | -| 1 |*`cudaD3D9DeviceListAll`* | | -| 2 |*`cudaD3D9DeviceListCurrentFrame`* | | -| 3 |*`cudaD3D9DeviceListNextFrame`* | | -| enum |***`cudaD3D9MapFlags`*** | | -| 0 |*`cudaD3D9MapFlagsNone`* | | -| 1 |*`cudaD3D9MapFlagsReadOnly`* | | -| 2 |*`cudaD3D9MapFlagsWriteDiscard`* | | -| enum |***`cudaD3D9RegisterFlags`*** | | -| 0 |*`cudaD3D9RegisterFlagsNone`* | | -| 1 |*`cudaD3D9RegisterFlagsArray`* | | -| enum |***`cudaD3D10DeviceList`*** | | -| 1 |*`cudaD3D10DeviceListAll`* | | -| 2 |*`cudaD3D10DeviceListCurrentFrame`* | | -| 3 |*`cudaD3D10DeviceListNextFrame`* | | -| enum |***`cudaD3D10MapFlags`*** | | -| 0 |*`cudaD3D10MapFlagsNone`* | | -| 1 |*`cudaD3D10MapFlagsReadOnly`* | | -| 2 |*`cudaD3D10MapFlagsWriteDiscard`* | | -| enum |***`cudaD3D10RegisterFlags`*** | | -| 0 |*`cudaD3D10RegisterFlagsNone`* | | -| 1 |*`cudaD3D10RegisterFlagsArray`* | | -| enum |***`cudaD3D11DeviceList`*** | | -| 1 |*`cudaD3D11DeviceListAll`* | | -| 2 |*`cudaD3D11DeviceListCurrentFrame`* | | -| 3 |*`cudaD3D11DeviceListNextFrame`* | | -| struct |`cudaArray` |`hipArray` | -| typedef |`cudaArray_t` |`hipArray_t` | -| typedef |`cudaArray_const_t` |`hipArray_const_t` | -| typedef |`cudaEvent_t` |`hipEvent_t` | -| struct |`CUevent_st` |`ihipEvent_t` | -| typedef |`cudaGraphicsResource_t` | | -| struct |`cudaMipmappedArray` |`hipMipmappedArray` | -| typedef |`cudaMipmappedArray_t` |`hipMipmappedArray_t` | -| typedef |`cudaMipmappedArray_const_t` |`hipMipmappedArray_const_t` | -| enum |***`cudaOutputMode`*** | | -| typedef |***`cudaOutputMode_t`*** | | -| 0x00 |*`cudaKeyValuePair`* | | -| 0x01 |*`cudaCSV`* | | -| typedef |`cudaStream_t` |`hipStream_t` | -| struct |`CUstream_st` |`ihipStream_t` | -| typedef |`cudaStreamCallback_t` |`hipStreamCallback_t` | -| typedef |`cudaSurfaceObject_t` |`hipSurfaceObject_t` | -| typedef |`cudaTextureObject_t` |`hipTextureObject_t` | -| typedef |`CUuuid_stcudaUUID_t` | | -| define |`CUDA_EGL_MAX_PLANES` | | -| define |`CUDA_IPC_HANDLE_SIZE` | | -| define |`cudaArrayColorAttachment` | | -| define |`cudaArrayCubemap` |`hipArrayCubemap` | -| define |`cudaArrayDefault` |`hipArrayDefault` | -| define |`cudaArrayLayered` |`hipArrayLayered` | -| define |`cudaArraySurfaceLoadStore` |`hipArraySurfaceLoadStore` | -| define |`cudaArrayTextureGather` |`hipArrayTextureGather` | -| define |`cudaCooperativeLaunchMultiDeviceNoPreSync` | | -| define |`cudaCooperativeLaunchMultiDeviceNoPostSync` | | -| define |`cudaCpuDeviceId` | | -| define |`cudaInvalidDeviceId` | | -| define |`cudaDeviceBlockingSync` |`hipDeviceScheduleBlockingSync` | -| define |`cudaDeviceLmemResizeToMax` |`hipDeviceLmemResizeToMax` | 0x16 | -| define |`cudaDeviceMapHost` |`hipDeviceMapHost` | -| define |`cudaDeviceMask` | | -| define |`cudaDevicePropDontCare` | | -| define |`cudaDeviceScheduleAuto` |`hipDeviceScheduleAuto` | -| define |`cudaDeviceScheduleBlockingSync` |`hipDeviceScheduleBlockingSync` | -| define |`cudaDeviceScheduleMask` |`hipDeviceScheduleMask` | -| define |`cudaDeviceScheduleSpin` |`hipDeviceScheduleSpin` | -| define |`cudaDeviceScheduleYield` |`hipDeviceScheduleYield` | -| define |`cudaEventDefault` |`hipEventDefault` | -| define |`cudaEventBlockingSync` |`hipEventBlockingSync` | -| define |`cudaEventDisableTiming` |`hipEventDisableTiming` | -| define |`cudaEventInterprocess` |`hipEventInterprocess` | -| define |`cudaHostAllocDefault` |`hipHostMallocDefault` | -| define |`cudaHostAllocMapped` |`hipHostMallocMapped` | -| define |`cudaHostAllocPortable` |`hipHostMallocPortable` | -| define |`cudaHostAllocWriteCombined` |`hipHostMallocWriteCombined` | -| define |`cudaHostRegisterDefault` |`hipHostRegisterDefault` | -| define |`cudaHostRegisterIoMemory` |`hipHostRegisterIoMemory` | -| define |`cudaHostRegisterMapped` |`hipHostRegisterMapped` | -| define |`cudaHostRegisterPortable` |`hipHostRegisterPortable` | -| define |`cudaIpcMemLazyEnablePeerAccess` |`hipIpcMemLazyEnablePeerAccess` | 0 | -| define |`cudaMemAttachGlobal` | | -| define |`cudaMemAttachHost` | | -| define |`cudaMemAttachSingle` | | -| define |`cudaOccupancyDefault` | | -| define |`cudaOccupancyDisableCachingOverride` | | -| define |`cudaPeerAccessDefault` | | -| define |`cudaStreamDefault` |`hipStreamDefault` | -| define |`cudaStreamNonBlocking` |`hipStreamNonBlocking` | -| define |`cudaStreamLegacy` | | -| define |`cudaStreamPerThread` | | -| define |`cudaTextureType1D` |`hipTextureType1D` | -| define |`cudaTextureType2D` |`hipTextureType2D` | -| define |`cudaTextureType3D` |`hipTextureType3D` | -| define |`cudaTextureTypeCubemap` |`hipTextureTypeCubemap` | -| define |`cudaTextureType1DLayered` |`hipTextureType1DLayered` | -| define |`cudaTextureType2DLayered` |`hipTextureType2DLayered` | -| define |`cudaTextureTypeCubemapLayered` |`hipTextureTypeCubemapLayered` | -| enum |***`cudaDataType_t`*** |***`hipblasDatatype_t`*** | -| enum |***`cudaDataType`*** |***`hipblasDatatype_t`*** | -| 2 |*`CUDA_R_16F`* |*`HIPBLAS_R_16F`* | 150 | -| 6 |*`CUDA_C_16F`* |*`HIPBLAS_C_16F`* | 153 | -| 0 |*`CUDA_R_32F`* |*`HIPBLAS_R_32F`* | 151 | -| 4 |*`CUDA_C_32F`* |*`HIPBLAS_C_32F`* | 154 | -| 1 |*`CUDA_R_64F`* |*`HIPBLAS_R_64F`* | 152 | -| 5 |*`CUDA_C_64F`* |*`HIPBLAS_C_64F`* | 155 | -| 3 |*`CUDA_R_8I`* | | -| 7 |*`CUDA_C_8I`* | | -| 8 |*`CUDA_R_8U`* | | -| 9 |*`CUDA_C_8U`* | | -| 10 |*`CUDA_R_32I`* | | -| 11 |*`CUDA_C_32I`* | | -| 12 |*`CUDA_R_32U`* | | -| 13 |*`CUDA_C_32U`* | | -| struct |`cudaExternalMemoryBufferDesc` | | -| struct |`cudaExternalMemoryHandleDesc` | | -| struct |`cudaExternalMemoryMipmappedArrayDesc` | | -| struct |`cudaExternalSemaphoreHandleDesc` | | -| struct |`cudaExternalSemaphoreSignalParams` | | -| struct |`cudaExternalSemaphoreWaitParams` | | -| struct |`cudaHostNodeParams` | | -| struct |`cudaLaunchParams` | | -| struct |`cudaMemsetParams` | | -| struct |`CUeglStreamConnection_st` | | -| typedef |`cudaEglStreamConnection` | | -| define |`cudaExternalMemoryDedicated` | | -| typedef |`cudaExternalMemory_t` | | -| struct |`CUexternalMemory_st` | | -| typedef |`cudaExternalSemaphore_t` | | -| struct |`CUexternalSemaphore_st` | | -| typedef |`cudaGraph_t` | | -| struct |`CUgraph_st` | | -| typedef |`cudaGraphNode_t` | | -| struct |`CUgraphNode_st` | | -| typedef |`cudaGraphExec_t` | | -| struct |`CUgraphExec_st` | | -| typedef |`cudaGraphicsResource_t` | | -| struct |`cudaGraphicsResource` | | -| typedef |`cudaHostFn_t` | | -| enum |***`libraryPropertyType`*** | | -| typedef |***`libraryPropertyType_t`*** | | -| 0 |*`MAJOR_VERSION`* | | -| 1 |*`MINOR_VERSION`* | | -| 2 |*`PATCH_LEVEL`* | | +| **type** | **CUDA** |**CUDA version\***| **HIP** |**HIP value** (if differs) | +|-------------:|-----------------------------------------------------|:----------------:|------------------------------------------------------------|---------------------------| +| struct |`cudaChannelFormatDesc` | |`hipChannelFormatDesc` | +| struct |`cudaDeviceProp` | |`hipDeviceProp_t` | +| struct |`cudaEglFrame` | 9.1 | | +| typedef |`cudaEglFrame_st` | 9.1 | | +| struct |`cudaEglPlaneDesc` | 9.1 | | +| typedef |`cudaEglPlaneDesc_st` | 9.1 | | +| struct |`cudaExtent` | |`hipExtent` | +| struct |`cudaFuncAttributes` | |`hipFuncAttributes` | +| struct |`cudaIpcEventHandle_t` | |`hipIpcEventHandle_t` | +| struct |`cudaIpcMemHandle_t` | |`hipIpcMemHandle_t` | +| struct |`cudaMemcpy3DParms` | |`hipMemcpy3DParms` | +| struct |`cudaMemcpy3DPeerParms` | | | +| struct |`cudaPitchedPtr` | |`hipPitchedPtr` | +| struct |`cudaPointerAttributes` | |`hipPointerAttribute_t` | +| struct |`cudaPos` | |`hipPos` | +| struct |`cudaResourceDesc` | |`hipResourceDesc` | +| struct |`cudaResourceViewDesc` | |`hipResourceViewDesc` | +| struct |`cudaTextureDesc` | |`hipTextureDesc` | +| struct |`textureReference` | |`textureReference` | +| struct |`surfaceReference` | | | +| struct |`CUuuid_st` | | | +| enum |***`cudaCGScope`*** | 9.0 | | +| 0 |*`cudaCGScopeInvalid`* | 9.0 | | +| 1 |*`cudaCGScopeGrid`* | 9.0 | | +| 2 |*`cudaCGScopeMultiGrid`* | 9.0 | | +| enum |***`cudaChannelFormatKind`*** | |***`hipChannelFormatKind`*** | +| 0 |*`cudaChannelFormatKindSigned`* | |*`hipChannelFormatKindSigned`* | +| 1 |*`cudaChannelFormatKindUnsigned`* | |*`hipChannelFormatKindUnsigned`* | +| 2 |*`cudaChannelFormatKindFloat`* | |*`hipChannelFormatKindFloat`* | +| 3 |*`cudaChannelFormatKindNone`* | |*`hipChannelFormatKindNone`* | +| enum |***`cudaComputeMode`*** | |***`hipComputeMode`*** | +| 0 |*`cudaComputeModeDefault`* | |*`hipComputeModeDefault`* | +| 1 |*`cudaComputeModeExclusive`* | |*`hipComputeModeExclusive`* | +| 2 |*`cudaComputeModeProhibited`* | |*`hipComputeModeProhibited`* | +| 3 |*`cudaComputeModeExclusiveProcess`* | |*`hipComputeModeExclusiveProcess`* | +| enum |***`cudaDeviceAttr`*** | |***`hipDeviceAttribute_t`*** | +| 1 |*`cudaDevAttrMaxThreadsPerBlock`* | |*`hipDeviceAttributeMaxThreadsPerBlock`* | +| 2 |*`cudaDevAttrMaxBlockDimX`* | |*`hipDeviceAttributeMaxBlockDimX`* | +| 3 |*`cudaDevAttrMaxBlockDimY`* | |*`hipDeviceAttributeMaxBlockDimY`* | +| 4 |*`cudaDevAttrMaxBlockDimZ`* | |*`hipDeviceAttributeMaxBlockDimZ`* | +| 5 |*`cudaDevAttrMaxGridDimX`* | |*`hipDeviceAttributeMaxGridDimX`* | +| 6 |*`cudaDevAttrMaxGridDimY`* | |*`hipDeviceAttributeMaxGridDimY`* | +| 7 |*`cudaDevAttrMaxGridDimZ`* | |*`hipDeviceAttributeMaxGridDimZ`* | +| 8 |*`cudaDevAttrMaxSharedMemoryPerBlock`* | |*`hipDeviceAttributeMaxSharedMemoryPerBlock`* | +| 9 |*`cudaDevAttrTotalConstantMemory`* | |*`hipDeviceAttributeTotalConstantMemory`* | +| 10 |*`cudaDevAttrWarpSize`* | |*`hipDeviceAttributeWarpSize`* | +| 11 |*`cudaDevAttrMaxPitch`* | | | +| 12 |*`cudaDevAttrMaxRegistersPerBlock`* | |*`hipDeviceAttributeMaxRegistersPerBlock`* | +| 13 |*`cudaDevAttrClockRate`* | |*`hipDeviceAttributeClockRate`* | +| 14 |*`cudaDevAttrTextureAlignment`* | | | +| 15 |*`cudaDevAttrGpuOverlap`* | | | +| 16 |*`cudaDevAttrMultiProcessorCount`* | |*`hipDeviceAttributeMultiprocessorCount`* | +| 17 |*`cudaDevAttrKernelExecTimeout`* | | | +| 18 |*`cudaDevAttrIntegrated`* | |*`hipDeviceAttributeIntegrated`* | +| 19 |*`cudaDevAttrCanMapHostMemory`* | | | +| 20 |*`cudaDevAttrComputeMode`* | |*`hipDeviceAttributeComputeMode`* | +| 21 |*`cudaDevAttrMaxTexture1DWidth`* | |*`hipDeviceAttributeMaxTexture1DWidth`* | +| 22 |*`cudaDevAttrMaxTexture2DWidth`* | |*`hipDeviceAttributeMaxTexture2DWidth`* | +| 23 |*`cudaDevAttrMaxTexture2DHeight`* | |*`hipDeviceAttributeMaxTexture2DHeight`* | +| 24 |*`cudaDevAttrMaxTexture3DWidth`* | |*`hipDeviceAttributeMaxTexture3DWidth`* | +| 25 |*`cudaDevAttrMaxTexture3DHeight`* | |*`hipDeviceAttributeMaxTexture3DHeight`* | +| 26 |*`cudaDevAttrMaxTexture3DDepth`* | |*`hipDeviceAttributeMaxTexture3DDepth`* | +| 27 |*`cudaDevAttrMaxTexture2DLayeredWidth`* | | | +| 28 |*`cudaDevAttrMaxTexture2DLayeredHeight`* | | | +| 29 |*`cudaDevAttrMaxTexture2DLayeredLayers`* | | | +| 30 |*`cudaDevAttrSurfaceAlignment`* | | | +| 31 |*`cudaDevAttrConcurrentKernels`* | |*`hipDeviceAttributeConcurrentKernels`* | +| 32 |*`cudaDevAttrEccEnabled`* | | | +| 33 |*`cudaDevAttrPciBusId`* | |*`hipDeviceAttributePciBusId`* | +| 34 |*`cudaDevAttrPciDeviceId`* | |*`hipDeviceAttributePciDeviceId`* | +| 35 |*`cudaDevAttrTccDriver`* | | | +| 36 |*`cudaDevAttrMemoryClockRate`* | |*`hipDeviceAttributeMemoryClockRate`* | +| 37 |*`cudaDevAttrGlobalMemoryBusWidth`* | |*`hipDeviceAttributeMemoryBusWidth`* | +| 38 |*`cudaDevAttrL2CacheSize`* | |*`hipDeviceAttributeL2CacheSize`* | +| 39 |*`cudaDevAttrMaxThreadsPerMultiProcessor`* | |*`hipDeviceAttributeMaxThreadsPerMultiProcessor`* | +| 40 |*`cudaDevAttrAsyncEngineCount`* | | | +| 41 |*`cudaDevAttrUnifiedAddressing`* | | | +| 42 |*`cudaDevAttrMaxTexture1DLayeredWidth`* | | | +| 43 |*`cudaDevAttrMaxTexture1DLayeredLayers`* | | | +| 44 | | | | +| 45 |*`cudaDevAttrMaxTexture2DGatherWidth`* | | | +| 46 |*`cudaDevAttrMaxTexture2DGatherHeight`* | | | +| 47 |*`cudaDevAttrMaxTexture3DWidthAlt`* | | | +| 48 |*`cudaDevAttrMaxTexture3DHeightAlt`* | | | +| 49 |*`cudaDevAttrMaxTexture3DDepthAlt`* | | | +| 50 |*`cudaDevAttrPciDomainId`* | | | +| 51 |*`cudaDevAttrTexturePitchAlignment`* | | | +| 52 |*`cudaDevAttrMaxTextureCubemapWidth`* | | | +| 53 |*`cudaDevAttrMaxTextureCubemapLayeredWidth`* | | | +| 54 |*`cudaDevAttrMaxTextureCubemapLayeredLayers`* | | | +| 55 |*`cudaDevAttrMaxSurface1DWidth`* | | | +| 56 |*`cudaDevAttrMaxSurface2DWidth`* | | | +| 57 |*`cudaDevAttrMaxSurface2DHeight`* | | | +| 58 |*`cudaDevAttrMaxSurface3DWidth`* | | | +| 59 |*`cudaDevAttrMaxSurface3DHeight`* | | | +| 60 |*`cudaDevAttrMaxSurface3DDepth`* | | | +| 61 |*`cudaDevAttrMaxSurface1DLayeredWidth`* | | | +| 62 |*`cudaDevAttrMaxSurface1DLayeredLayers`* | | | +| 63 |*`cudaDevAttrMaxSurface2DLayeredWidth`* | | | +| 64 |*`cudaDevAttrMaxSurface2DLayeredHeight`* | | | +| 65 |*`cudaDevAttrMaxSurface2DLayeredLayers`* | | | +| 66 |*`cudaDevAttrMaxSurfaceCubemapWidth`* | | | +| 67 |*`cudaDevAttrMaxSurfaceCubemapLayeredWidth`* | | | +| 68 |*`cudaDevAttrMaxSurfaceCubemapLayeredLayers`* | | | +| 69 |*`cudaDevAttrMaxTexture1DLinearWidth`* | | | +| 70 |*`cudaDevAttrMaxTexture2DLinearWidth`* | | | +| 71 |*`cudaDevAttrMaxTexture2DLinearHeight`* | | | +| 72 |*`cudaDevAttrMaxTexture2DLinearPitch`* | | | +| 73 |*`cudaDevAttrMaxTexture2DMipmappedWidth* | | | +| 74 |*`cudaDevAttrMaxTexture2DMipmappedHeight`* | | | +| 75 |*`cudaDevAttrComputeCapabilityMajor`* | |*`hipDeviceAttributeComputeCapabilityMajor`* | +| 76 |*`cudaDevAttrComputeCapabilityMinor`* | |*`hipDeviceAttributeComputeCapabilityMinor`* | +| 77 |*`cudaDevAttrMaxTexture1DMipmappedWidth`* | | | +| 78 |*`cudaDevAttrStreamPrioritiesSupported`* | | | +| 79 |*`cudaDevAttrGlobalL1CacheSupported`* | | | +| 80 |*`cudaDevAttrLocalL1CacheSupported`* | | | +| 81 |*`cudaDevAttrMaxSharedMemoryPerMultiprocessor`* | |*`hipDeviceAttributeMaxSharedMemoryPerMultiprocessor`* | +| 82 |*`cudaDevAttrMaxRegistersPerMultiprocessor`* | | | +| 83 |*`cudaDevAttrManagedMemory`* | | | +| 84 |*`cudaDevAttrIsMultiGpuBoard`* | |*`hipDeviceAttributeIsMultiGpuBoard`* | +| 85 |*`cudaDevAttrMultiGpuBoardGroupID`* | | | +| 86 |*`cudaDevAttrHostNativeAtomicSupported`* | 8.0 | | +| 87 |*`cudaDevAttrSingleToDoublePrecisionPerfRatio`* | 8.0 | | +| 88 |*`cudaDevAttrPageableMemoryAccess`* | 8.0 | | +| 89 |*`cudaDevAttrConcurrentManagedAccess`* | 8.0 | | +| 90 |*`cudaDevAttrComputePreemptionSupported`* | 8.0 | | +| 91 |*`cudaDevAttrCanUseHostPointerForRegisteredMem`* | 8.0 | | +| 92 |*`cudaDevAttrReserved92`* | 9.0 | | +| 93 |*`cudaDevAttrReserved93`* | 9.0 | | +| 94 |*`cudaDevAttrReserved94`* | 9.0 | | +| 95 |*`cudaDevAttrCooperativeLaunch`* | 9.0 | | +| 96 |*`cudaDevAttrCooperativeMultiDeviceLaunch`* | 9.0 | | +| 97 |*`cudaDevAttrMaxSharedMemoryPerBlockOptin`* | 9.0 | | +| 98 |*`cudaDevAttrCanFlushRemoteWrites`* | 9.2 | | +| 99 |*`cudaDevAttrHostRegisterSupported`* | 9.2 | | +| 100 |*`cudaDevAttrPageableMemoryAccessUsesHostPageTables`*| 9.2 | | +| 101 |*`cudaDevAttrDirectManagedMemAccessFromHost`* | 9.2 | | +| enum |***`cudaDeviceP2PAttr`*** | 8.0 | | +| 1 |*`cudaDevP2PAttrPerformanceRank`* | 8.0 | | +| 2 |*`cudaDevP2PAttrAccessSupported`* | 8.0 | | +| 3 |*`cudaDevP2PAttrNativeAtomicSupported`* | 8.0 | | +| 4 |*`cudaDevP2PAttrCudaArrayAccessSupported`* | 9.2 | | +| enum |***`cudaEglColorFormat`*** | 9.1 | | +| 0 |*`cudaEglColorFormatYUV420Planar`* | 9.1 | | +| 1 |*`cudaEglColorFormatYUV420SemiPlanar`* | 9.1 | | +| 2 |*`cudaEglColorFormatYUV422Planar`* | 9.1 | | +| 3 |*`cudaEglColorFormatYUV422SemiPlanar`* | 9.1 | | +| 4 |*`cudaEglColorFormatRGB`* | 9.1 | | +| 5 |*`cudaEglColorFormatBGR`* | 9.1 | | +| 6 |*`cudaEglColorFormatARGB`* | 9.1 | | +| 7 |*`cudaEglColorFormatRGBA`* | 9.1 | | +| 8 |*`cudaEglColorFormatL`* | 9.1 | | +| 9 |*`cudaEglColorFormatR`* | 9.1 | | +| 10 |*`cudaEglColorFormatYUV444Planar`* | 9.1 | | +| 11 |*`cudaEglColorFormatYUV444SemiPlanar`* | 9.1 | | +| 12 |*`cudaEglColorFormatYUYV422`* | 9.1 | | +| 13 |*`cudaEglColorFormatUYVY422`* | 9.1 | | +| 14 |*`cudaEglColorFormatABGR`* | 9.1 | | +| 15 |*`cudaEglColorFormatBGRA`* | 9.1 | | +| 16 |*`cudaEglColorFormatA`* | 9.1 | | +| 17 |*`cudaEglColorFormatRG`* | 9.1 | | +| 18 |*`cudaEglColorFormatAYUV`* | 9.1 | | +| 19 |*`cudaEglColorFormatYVU444SemiPlanar`* | 9.1 | | +| 20 |*`cudaEglColorFormatYVU422SemiPlanar`* | 9.1 | | +| 21 |*`cudaEglColorFormatYVU420SemiPlanar`* | 9.1 | | +| 22 |*`cudaEglColorFormatY10V10U10_444SemiPlanar`* | 9.1 | | +| 23 |*`cudaEglColorFormatY10V10U10_420SemiPlanar`* | 9.1 | | +| 24 |*`cudaEglColorFormatY12V12U12_444SemiPlanar`* | 9.1 | | +| 25 |*`cudaEglColorFormatY12V12U12_420SemiPlanar`* | 9.1 | | +| 26 |*`cudaEglColorFormatVYUY_ER`* | 9.1 | | +| 27 |*`cudaEglColorFormatUYVY_ER`* | 9.1 | | +| 28 |*`cudaEglColorFormatYUYV_ER`* | 9.1 | | +| 29 |*`cudaEglColorFormatYVYU_ER`* | 9.1 | | +| 30 |*`cudaEglColorFormatYUV_ER`* | 9.1 | | +| 31 |*`cudaEglColorFormatYUVA_ER`* | 9.1 | | +| 32 |*`cudaEglColorFormatAYUV_ER`* | 9.1 | | +| 33 |*`cudaEglColorFormatYUV444Planar_ER`* | 9.1 | | +| 34 |*`cudaEglColorFormatYUV422Planar_ER`* | 9.1 | | +| 35 |*`cudaEglColorFormatYUV420Planar_ER`* | 9.1 | | +| 36 |*`cudaEglColorFormatYUV444SemiPlanar_ER`* | 9.1 | | +| 37 |*`cudaEglColorFormatYUV422SemiPlanar_ER`* | 9.1 | | +| 38 |*`cudaEglColorFormatYUV420SemiPlanar_ER`* | 9.1 | | +| 39 |*`cudaEglColorFormatYVU444Planar_ER`* | 9.1 | | +| 40 |*`cudaEglColorFormatYVU422Planar_ER`* | 9.1 | | +| 41 |*`cudaEglColorFormatYVU420Planar_ER`* | 9.1 | | +| 42 |*`cudaEglColorFormatYVU444SemiPlanar_ER`* | 9.1 | | +| 43 |*`cudaEglColorFormatYVU422SemiPlanar_ER`* | 9.1 | | +| 44 |*`cudaEglColorFormatYVU420SemiPlanar_ER`* | 9.1 | | +| 45 |*`cudaEglColorFormatBayerRGGB`* | 9.1 | | +| 46 |*`cudaEglColorFormatBayerBGGR`* | 9.1 | | +| 47 |*`cudaEglColorFormatBayerGRBG`* | 9.1 | | +| 48 |*`cudaEglColorFormatBayerGBRG`* | 9.1 | | +| 49 |*`cudaEglColorFormatBayer10RGGB`* | 9.1 | | +| 50 |*`cudaEglColorFormatBayer10BGGR`* | 9.1 | | +| 51 |*`cudaEglColorFormatBayer10GRBG`* | 9.1 | | +| 52 |*`cudaEglColorFormatBayer10GBRG`* | 9.1 | | +| 53 |*`cudaEglColorFormatBayer12RGGB`* | 9.1 | | +| 54 |*`cudaEglColorFormatBayer12BGGR`* | 9.1 | | +| 55 |*`cudaEglColorFormatBayer12GRBG`* | 9.1 | | +| 56 |*`cudaEglColorFormatBayer12GBRG`* | 9.1 | | +| 57 |*`cudaEglColorFormatBayer14RGGB`* | 9.1 | | +| 58 |*`cudaEglColorFormatBayer14BGGR`* | 9.1 | | +| 59 |*`cudaEglColorFormatBayer14GRBG`* | 9.1 | | +| 60 |*`cudaEglColorFormatBayer14GBRG`* | 9.1 | | +| 61 |*`cudaEglColorFormatBayer20RGGB`* | 9.1 | | +| 62 |*`cudaEglColorFormatBayer20BGGR`* | 9.1 | | +| 63 |*`cudaEglColorFormatBayer20GRBG`* | 9.1 | | +| 64 |*`cudaEglColorFormatBayer20GBRG`* | 9.1 | | +| 65 |*`cudaEglColorFormatYVU444Planar`* | 9.1 | | +| 66 |*`cudaEglColorFormatYVU422Planar`* | 9.1 | | +| 67 |*`cudaEglColorFormatYVU420Planar`* | 9.1 | | +| 68 |*`cudaEglColorFormatBayerIspRGGB`* | 9.2 | | +| 69 |*`cudaEglColorFormatBayerIspBGGR`* | 9.2 | | +| 70 |*`cudaEglColorFormatBayerIspGRBG`* | 9.2 | | +| 71 |*`cudaEglColorFormatBayerIspGBRG`* | 9.2 | | +| enum |***`cudaEglFrameType`*** | 9.1 | | +| 0 |*`cudaEglFrameTypeArray`* | 9.1 | | +| 1 |*`cudaEglFrameTypePitch`* | 9.1 | | +| enum |***`cudaExternalMemoryHandleType`*** | 10.0 | | +| 1 |*`cudaExternalMemoryHandleTypeOpaqueFd`* | 10.0 | | +| 2 |*`cudaExternalMemoryHandleTypeOpaqueWin32`* | 10.0 | | +| 3 |*`cudaExternalMemoryHandleTypeOpaqueWin32Kmt`* | 10.0 | | +| 4 |*`cudaExternalMemoryHandleTypeD3D12Heap`* | 10.0 | | +| 5 |*`cudaExternalMemoryHandleTypeD3D12Resource`* | 10.0 | | +| enum |***`cudaExternalSemaphoreHandleType`*** | 10.0 | | +| 1 |*`cudaExternalSemaphoreHandleTypeOpaqueFd`* | 10.0 | | +| 2 |*`cudaExternalSemaphoreHandleTypeOpaqueWin32`* | 10.0 | | +| 3 |*`cudaExternalSemaphoreHandleTypeOpaqueWin32Kmt`* | 10.0 | | +| 4 |*`cudaExternalSemaphoreHandleTypeD3D12Fence`* | 10.0 | | +| enum |***`cudaFuncAttribute`*** | 9.0 | | +| 8 |*`cudaFuncAttributeMaxDynamicSharedMemorySize`* | 9.0 | | +| 9 |*`cudaFuncAttributePreferredSharedMemoryCarveout`* | 9.0 | | +| 10 |*`cudaFuncAttributeMax`* | 9.0 | | +| enum |***`cudaEglResourceLocationFlags`*** | 9.1 | | +| 0x00 |*`cudaEglResourceLocationSysmem`* | 9.1 | | +| 0x01 |*`cudaEglResourceLocationVidmem`* | 9.1 | | +| enum |***`cudaError`*** | |***`hipError_t`*** | +| typedef |***`cudaError_t`*** | |***`hipError_t`*** | +| 0 |*`cudaSuccess`* | |*`hipSuccess`* | +| 1 |*`cudaErrorMissingConfiguration`* | |*`hipErrorMissingConfiguration`* | 1001 | +| 2 |*`cudaErrorMemoryAllocation`* | |*`hipErrorMemoryAllocation`* | 1002 | +| 3 |*`cudaErrorInitializationError`* | |*`hipErrorInitializationError`* | 1003 | +| 4 |*`cudaErrorLaunchFailure`* | |*`hipErrorLaunchFailure`* | 1004 | +| 5 |*`cudaErrorPriorLaunchFailure`* | |*`hipErrorPriorLaunchFailure`* | 1005 | +| 6 |*`cudaErrorLaunchTimeout`* | |*`hipErrorLaunchTimeOut`* | 1006 | +| 7 |*`cudaErrorLaunchOutOfResources`* | |*`hipErrorLaunchOutOfResources`* | 1007 | +| 8 |*`cudaErrorInvalidDeviceFunction`* | |*`hipErrorInvalidDeviceFunction`* | 1008 | +| 9 |*`cudaErrorInvalidConfiguration`* | |*`hipErrorInvalidConfiguration`* | 1009 | +| 10 |*`cudaErrorInvalidDevice`* | |*`hipErrorInvalidDevice`* | 1010 | +| 11 |*`cudaErrorInvalidValue`* | |*`hipErrorInvalidValue`* | 1011 | +| 12 |*`cudaErrorInvalidPitchValue`* | | | +| 13 |*`cudaErrorInvalidSymbol`* | |*`hipErrorInvalidSymbol`* | 701 | +| 14 |*`cudaErrorMapBufferObjectFailed`* | |*`hipErrorMapFailed`* | 205 | +| 15 |*`cudaErrorUnmapBufferObjectFailed`* | |*`hipErrorUnmapFailed`* | 206 | +| 16 |*`cudaErrorInvalidHostPointer`* | | | +| 17 |*`cudaErrorInvalidDevicePointer`* | |*`hipErrorInvalidDevicePointer`* | 1017 | +| 18 |*`cudaErrorInvalidTexture`* | | | +| 19 |*`cudaErrorInvalidTextureBinding`* | | | +| 20 |*`cudaErrorInvalidChannelDescriptor`* | | | +| 21 |*`cudaErrorInvalidMemcpyDirection`* | |*`hipErrorInvalidMemcpyDirection`* | 1021 | +| 22 |*`cudaErrorAddressOfConstant`* | | | +| 23 |*`cudaErrorTextureFetchFailed`* | | | +| 24 |*`cudaErrorTextureNotBound`* | | | +| 25 |*`cudaErrorSynchronizationError`* | | | +| 26 |*`cudaErrorInvalidFilterSetting`* | | | +| 27 |*`cudaErrorInvalidNormSetting`* | | | +| 28 |*`cudaErrorMixedDeviceExecution`* | | | +| 29 |*`cudaErrorCudartUnloading`* | |*`hipErrorDeinitialized`* | 4 | +| 30 |*`cudaErrorUnknown`* | |*`hipErrorUnknown`* | 1030 | +| 31 |*`cudaErrorNotYetImplemented`* | | | +| 32 |*`cudaErrorMemoryValueTooLarge`* | | | +| 33 |*`cudaErrorInvalidResourceHandle`* | |*`hipErrorInvalidResourceHandle`* | 1033 | +| 34 |*`cudaErrorNotReady`* | |*`hipErrorNotReady`* | 1034 | +| 35 |*`cudaErrorInsufficientDriver`* | |*`hipErrorInsufficientDriver`* | +| 36 |*`cudaErrorSetOnActiveProcess`* | | | +| 37 |*`cudaErrorInvalidSurface`* | | | +| 38 |*`cudaErrorNoDevice`* | |*`hipErrorNoDevice`* | 1038 | +| 39 |*`cudaErrorECCUncorrectable`* | |*`hipErrorECCNotCorrectable`* | 214 | +| 40 |*`cudaErrorSharedObjectSymbolNotFound`* | |*`hipErrorSharedObjectSymbolNotFound`* | 302 | +| 41 |*`cudaErrorSharedObjectInitFailed`* | |*`hipErrorSharedObjectInitFailed`* | 303 | +| 42 |*`cudaErrorUnsupportedLimit`* | |*`hipErrorUnsupportedLimit`* | 215 | +| 43 |*`cudaErrorDuplicateVariableName`* | | | +| 44 |*`cudaErrorDuplicateTextureName`* | | | +| 45 |*`cudaErrorDuplicateSurfaceName`* | | | +| 46 |*`cudaErrorDevicesUnavailable`* | | | +| 47 |*`cudaErrorInvalidKernelImage`* | |*`hipErrorInvalidImage`* | 200 | +| 48 |*`cudaErrorNoKernelImageForDevice`* | |*`hipErrorNoBinaryForGpu`* | 209 | +| 49 |*`cudaErrorIncompatibleDriverContext`* | | | +| 50 |*`cudaErrorPeerAccessAlreadyEnabled`* | |*`hipErrorPeerAccessAlreadyEnabled`* | 1050 | +| 51 |*`cudaErrorPeerAccessNotEnabled`* | |*`hipErrorPeerAccessNotEnabled`* | 1051 | +| 52 | | | | +| 53 | | | | +| 54 |*`cudaErrorDeviceAlreadyInUse`* | | | +| 55 |*`cudaErrorProfilerDisabled`* | |*`hipErrorProfilerDisabled`* | 5 | +| 56 |*`cudaErrorProfilerNotInitialized`* | |*`hipErrorProfilerNotInitialized`* | 6 | +| 57 |*`cudaErrorProfilerAlreadyStarted`* | |*`hipErrorProfilerAlreadyStarted`* | 7 | +| 58 |*`cudaErrorProfilerAlreadyStopped`* | |*`hipErrorProfilerAlreadyStopped`* | 8 | +| 59 |*`cudaErrorAssert`* | |*`hipErrorAssert`* | 1081 | +| 60 |*`cudaErrorTooManyPeers`* | | | +| 61 |*`cudaErrorHostMemoryAlreadyRegistered`* | |*`hipErrorHostMemoryAlreadyRegistered`* | 1061 | +| 62 |*`cudaErrorHostMemoryNotRegistered`* | |*`hipErrorHostMemoryNotRegistered`* | 1062 | +| 63 |*`cudaErrorOperatingSystem`* | |*`hipErrorOperatingSystem`* | 304 | +| 64 |*`cudaErrorPeerAccessUnsupported`* | |*`hipErrorPeerAccessUnsupported`* | 217 | +| 65 |*`cudaErrorLaunchMaxDepthExceeded`* | | | +| 66 |*`cudaErrorLaunchFileScopedTex`* | | | +| 67 |*`cudaErrorLaunchFileScopedSurf`* | | | +| 68 |*`cudaErrorSyncDepthExceeded`* | | | +| 69 |*`cudaErrorLaunchPendingCountExceeded`* | | | +| 70 |*`cudaErrorNotPermitted`* | | | +| 71 |*`cudaErrorNotSupported`* | | | +| 72 |*`cudaErrorHardwareStackError`* | | | +| 73 |*`cudaErrorIllegalInstruction`* | | | +| 74 |*`cudaErrorMisalignedAddress`* | | | +| 75 |*`cudaErrorInvalidAddressSpace`* | | | +| 76 |*`cudaErrorInvalidPc`* | | | +| 77 |*`cudaErrorIllegalAddress`* | |*`hipErrorIllegalAddress`* | 700 | +| 78 |*`cudaErrorInvalidPtx`* | |*`hipErrorInvalidKernelFile`* | 218 | +| 79 |*`cudaErrorInvalidGraphicsContext`* | |*`hipErrorInvalidGraphicsContext`* | 219 | +| 80 |*`cudaErrorNvlinkUncorrectable`* | 8.0 | | +| 81 |*`cudaErrorJitCompilerNotFound`* | 9.0 | | +| 82 |*`cudaErrorCooperativeLaunchTooLarge`* | 9.0 | | +| 83 |*`cudaErrorSystemNotReady`* | 10.0 | | +| 84 |*`cudaErrorIllegalState`* | 10.0 | | +| 127 |*`cudaErrorStartupFailure`* | 10.0 | | +| 803 |*`cudaErrorSystemDriverMismatch`* | 10.0 | | +| 804 |*`cudaErrorCompatNotSupportedOnDevice`* | 10.0 | | +| 900 |*`cudaErrorStreamCaptureUnsupported`* | 10.0 | | +| 901 |*`cudaErrorStreamCaptureInvalidated`* | 10.0 | | +| 902 |*`cudaErrorStreamCaptureMerge`* | 10.0 | | +| 903 |*`cudaErrorStreamCaptureUnmatched`* | 10.0 | | +| 904 |*`cudaErrorStreamCaptureUnjoined`* | 10.0 | | +| 905 |*`cudaErrorStreamCaptureIsolation`* | 10.0 | | +| 906 |*`cudaErrorStreamCaptureImplicit`* | 10.0 | | +| 907 |*`cudaErrorCapturedEvent`* | 10.0 | | +| 908 |*`cudaErrorStreamCaptureWrongThread`* | 10.1 | | +| 10000 |*`cudaErrorApiFailureBase`* | | | +| enum |***`cudaFuncCache`*** | |***`hipFuncCache_t`*** | +| 0 |*`cudaFuncCachePreferNone`* | |*`hipFuncCachePreferNone`* | +| 1 |*`cudaFuncCachePreferShared`* | |*`hipFuncCachePreferShared`* | +| 2 |*`cudaFuncCachePreferL1`* | |*`hipFuncCachePreferL1`* | +| 3 |*`cudaFuncCachePreferEqual`* | |*`hipFuncCachePreferEqual`* | +| enum |***`cudaGraphicsCubeFace`*** | | | +| 0x00 |*`cudaGraphicsCubeFacePositiveX`* | | | +| 0x01 |*`cudaGraphicsCubeFaceNegativeX`* | | | +| 0x02 |*`cudaGraphicsCubeFacePositiveY`* | | | +| 0x03 |*`cudaGraphicsCubeFaceNegativeY`* | | | +| 0x04 |*`cudaGraphicsCubeFacePositiveZ`* | | | +| 0x05 |*`cudaGraphicsCubeFaceNegativeZ`* | | | +| enum |***`cudaGraphicsMapFlags`*** | | | +| 0 |*`cudaGraphicsMapFlagsNone`* | | | +| 1 |*`cudaGraphicsMapFlagsReadOnly`* | | | +| 2 |*`cudaGraphicsMapFlagsWriteDiscard`* | | | +| enum |***`cudaGraphicsRegisterFlags`*** | | | +| 0 |*`cudaGraphicsRegisterFlagsNone`* | | | +| 1 |*`cudaGraphicsRegisterFlagsReadOnly`* | | | +| 2 |*`cudaGraphicsRegisterFlagsWriteDiscard`* | | | +| 4 |*`cudaGraphicsRegisterFlagsSurfaceLoadStore`* | | | +| 8 |*`cudaGraphicsRegisterFlagsTextureGather`* | | | +| enum |***`cudaGraphNodeType`*** | 10.0 | | +| 0x00 |*`cudaGraphNodeTypeKernel`* | 10.0 | | +| 0x01 |*`cudaGraphNodeTypeMemcpy`* | 10.0 | | +| 0x02 |*`cudaGraphNodeTypeMemset`* | 10.0 | | +| 0x03 |*`cudaGraphNodeTypeHost`* | 10.0 | | +| 0x04 |*`cudaGraphNodeTypeGraph`* | 10.0 | | +| 0x05 |*`cudaGraphNodeTypeEmpty`* | 10.0 | | +| |*`cudaGraphNodeTypeCount`* | 10.0 | | +| enum |***`cudaLimit`*** | |***`hipLimit_t`*** | +| 0x00 |*`cudaLimitStackSize`* | | | +| 0x01 |*`cudaLimitPrintfFifoSize`* | | | +| 0x02 |*`cudaLimitMallocHeapSize`* | |*`hipLimitMallocHeapSize`* | +| 0x03 |*`cudaLimitDevRuntimeSyncDepth`* | | | +| 0x04 |*`cudaLimitDevRuntimePendingLaunchCount`* | | | +| 0x05 |*`cudaLimitMaxL2FetchGranularity`* | 10.0 | | +| enum |***`cudaMemcpyKind`*** | |***`hipMemcpyKind`*** | +| 0 |*`cudaMemcpyHostToHost`* | |*`hipMemcpyHostToHost`* | +| 1 |*`cudaMemcpyHostToDevice`* | |*`hipMemcpyHostToDevice`* | +| 2 |*`cudaMemcpyDeviceToHost`* | |*`hipMemcpyDeviceToHost`* | +| 3 |*`cudaMemcpyDeviceToDevice`* | |*`hipMemcpyDeviceToDevice`* | +| 4 |*`cudaMemcpyDefault`* | |*`hipMemcpyDefault`* | +| enum |***`cudaMemoryAdvise`*** | 8.0 | | +| 1 |*`cudaMemAdviseSetReadMostly`* | 8.0 | | +| 2 |*`cudaMemAdviseUnsetReadMostly`* | 8.0 | | +| 3 |*`cudaMemAdviseSetPreferredLocation`* | 8.0 | | +| 4 |*`cudaMemAdviseUnsetPreferredLocation`* | 8.0 | | +| 5 |*`cudaMemAdviseSetAccessedBy`* | 8.0 | | +| 6 |*`cudaMemAdviseUnsetAccessedBy`* | 8.0 | | +| enum |***`cudaMemoryType`*** | | | +| 0 |*`cudaMemoryTypeUnregistered`* | | | +| 1 |*`cudaMemoryTypeHost`* | | | +| 2 |*`cudaMemoryTypeDevice`* | | | +| 3 |*`cudaMemoryTypeManaged`* | 10.0 | | +| enum |***`cudaMemRangeAttribute`*** | 8.0 | | +| 1 |*`cudaMemRangeAttributeReadMostly`* | 8.0 | | +| 2 |*`cudaMemRangeAttributePreferredLocation`* | 8.0 | | +| 3 |*`cudaMemRangeAttributeAccessedBy`* | 8.0 | | +| 4 |*`cudaMemRangeAttributeLastPrefetchLocation`* | 8.0 | | +| enum |***`cudaResourceType`*** | |***`hipResourceType`*** | +| 0x00 |*`cudaResourceTypeArray`* | |*`hipResourceTypeArray`* | +| 0x01 |*`cudaResourceTypeMipmappedArray`* | |*`hipResourceTypeMipmappedArray`* | +| 0x02 |*`cudaResourceTypeLinear`* | |*`hipResourceTypeLinear`* | +| 0x03 |*`cudaResourceTypePitch2D`* | |*`hipResourceTypePitch2D`* | +| enum |***`cudaResourceViewFormat`*** | |***`hipResourceViewFormat`*** | +| 0x00 |*`cudaResViewFormatNone`* | |*`hipResViewFormatNone`* | +| 0x01 |*`cudaResViewFormatUnsignedChar1`* | |*`hipResViewFormatUnsignedChar1`* | +| 0x02 |*`cudaResViewFormatUnsignedChar2`* | |*`hipResViewFormatUnsignedChar2`* | +| 0x03 |*`cudaResViewFormatUnsignedChar4`* | |*`hipResViewFormatUnsignedChar4`* | +| 0x04 |*`cudaResViewFormatSignedChar1`* | |*`hipResViewFormatSignedChar1`* | +| 0x05 |*`cudaResViewFormatSignedChar2`* | |*`hipResViewFormatSignedChar2`* | +| 0x06 |*`cudaResViewFormatSignedChar4`* | |*`hipResViewFormatSignedChar4`* | +| 0x07 |*`cudaResViewFormatUnsignedShort1`* | |*`hipResViewFormatUnsignedShort1`* | +| 0x08 |*`cudaResViewFormatUnsignedShort2`* | |*`hipResViewFormatUnsignedShort2`* | +| 0x09 |*`cudaResViewFormatUnsignedShort4`* | |*`hipResViewFormatUnsignedShort4`* | +| 0x0a |*`cudaResViewFormatSignedShort1`* | |*`hipResViewFormatSignedShort1`* | +| 0x0b |*`cudaResViewFormatSignedShort2`* | |*`hipResViewFormatSignedShort2`* | +| 0x0c |*`cudaResViewFormatSignedShort4`* | |*`hipResViewFormatSignedShort4`* | +| 0x0d |*`cudaResViewFormatUnsignedInt1`* | |*`hipResViewFormatUnsignedInt1`* | +| 0x0e |*`cudaResViewFormatUnsignedInt2`* | |*`hipResViewFormatUnsignedInt2`* | +| 0x0f |*`cudaResViewFormatUnsignedInt4`* | |*`hipResViewFormatUnsignedInt4`* | +| 0x10 |*`cudaResViewFormatSignedInt1`* | |*`hipResViewFormatSignedInt1`* | +| 0x11 |*`cudaResViewFormatSignedInt2`* | |*`hipResViewFormatSignedInt2`* | +| 0x12 |*`cudaResViewFormatSignedInt4`* | |*`hipResViewFormatSignedInt4`* | +| 0x13 |*`cudaResViewFormatHalf1`* | |*`hipResViewFormatHalf1`* | +| 0x14 |*`cudaResViewFormatHalf2`* | |*`hipResViewFormatHalf2`* | +| 0x15 |*`cudaResViewFormatHalf4`* | |*`hipResViewFormatHalf4`* | +| 0x16 |*`cudaResViewFormatFloat1`* | |*`hipResViewFormatFloat1`* | +| 0x17 |*`cudaResViewFormatFloat2`* | |*`hipResViewFormatFloat2`* | +| 0x18 |*`cudaResViewFormatFloat4`* | |*`hipResViewFormatFloat4`* | +| 0x19 |*`cudaResViewFormatUnsignedBlockCompressed1`* | |*`hipResViewFormatUnsignedBlockCompressed1`* | +| 0x1a |*`cudaResViewFormatUnsignedBlockCompressed2`* | |*`hipResViewFormatUnsignedBlockCompressed2`* | +| 0x1b |*`cudaResViewFormatUnsignedBlockCompressed3`* | |*`hipResViewFormatUnsignedBlockCompressed3`* | +| 0x1c |*`cudaResViewFormatUnsignedBlockCompressed4`* | |*`hipResViewFormatUnsignedBlockCompressed4`* | +| 0x1d |*`cudaResViewFormatSignedBlockCompressed4`* | |*`hipResViewFormatSignedBlockCompressed4`* | +| 0x1e |*`cudaResViewFormatUnsignedBlockCompressed5`* | |*`hipResViewFormatUnsignedBlockCompressed5`* | +| 0x1f |*`cudaResViewFormatSignedBlockCompressed5`* | |*`hipResViewFormatSignedBlockCompressed5`* | +| 0x20 |*`cudaResViewFormatUnsignedBlockCompressed6H`* | |*`hipResViewFormatUnsignedBlockCompressed6H`* | +| 0x21 |*`cudaResViewFormatSignedBlockCompressed6H`* | |*`hipResViewFormatSignedBlockCompressed6H`* | +| 0x22 |*`cudaResViewFormatUnsignedBlockCompressed7`* | |*`hipResViewFormatUnsignedBlockCompressed7`* | +| enum |***`cudaSharedMemConfig`*** | |***`hipSharedMemConfig`*** | +| 0 |*`cudaSharedMemBankSizeDefault`* | |*`hipSharedMemBankSizeDefault`* | +| 1 |*`cudaSharedMemBankSizeFourByte`* | |*`hipSharedMemBankSizeFourByte`* | +| 2 |*`cudaSharedMemBankSizeEightByte`* | |*`hipSharedMemBankSizeEightByte`* | +| enum |***`cudaSharedCarveout`*** | 9.0 | | +| -1 |*`cudaSharedmemCarveoutDefault`* | 9.0 | | +| 100 |*`cudaSharedmemCarveoutMaxShared`* | 9.0 | | +| 0 |*`cudaSharedmemCarveoutMaxL1`* | 9.0 | | +| enum |***`cudaStreamCaptureStatus`*** | 10.0 | | +| 0 |*`cudaStreamCaptureStatusNone`* | 10.0 | | +| 1 |*`cudaStreamCaptureStatusActive`* | 10.0 | | +| 2 |*`cudaStreamCaptureStatusInvalidated`* | 10.0 | | +| enum |***`cudaStreamCaptureMode`*** | 10.1 | | +| 0 |*`cudaStreamCaptureModeGlobal`* | 10.1 | | +| 1 |*`cudaStreamCaptureModeThreadLocal`* | 10.1 | | +| 2 |*`cudaStreamCaptureModeRelaxed`* | 10.1 | | +| enum |***`cudaSurfaceBoundaryMode`*** | |***`hipSurfaceBoundaryMode`*** | +| 0 |*`cudaBoundaryModeZero`* | |*`hipBoundaryModeZero`* | +| 1 |*`cudaBoundaryModeClamp`* | |*`hipBoundaryModeClamp`* | +| 2 |*`cudaBoundaryModeTrap`* | |*`hipBoundaryModeTrap`* | +| enum |***`cudaSurfaceFormatMode`*** | | | +| 0 |*`cudaFormatModeForced`* | | | +| 1 |*`cudaFormatModeAuto`* | | | +| enum |***`cudaTextureAddressMode`*** | |***`hipTextureAddressMode`*** | +| 0 |*`cudaAddressModeWrap`* | |*`hipAddressModeWrap`* | +| 1 |*`cudaAddressModeClamp`* | |*`hipAddressModeClamp`* | +| 2 |*`cudaAddressModeMirror`* | |*`hipAddressModeMirror`* | +| 3 |*`cudaAddressModeBorder`* | |*`hipAddressModeBorder`* | +| enum |***`cudaTextureFilterMode`*** | |***`hipTextureFilterMode`*** | +| 0 |*`cudaFilterModePoint`* | |*`hipFilterModePoint`* | +| 1 |*`cudaFilterModeLinear`* | |*`hipFilterModeLinear`* | +| enum |***`cudaTextureReadMode`*** | |***`hipTextureReadMode`*** | +| 0 |*`cudaReadModeElementType`* | |*`hipReadModeElementType`* | +| 1 |*`cudaReadModeNormalizedFloat`* | |*`hipReadModeNormalizedFloat`* | +| enum |***`cudaGLDeviceList`*** | | | +| 1 |*`cudaGLDeviceListAll`* | | | +| 2 |*`cudaGLDeviceListCurrentFrame`* | | | +| 3 |*`cudaGLDeviceListNextFrame`* | | | +| enum |***`cudaGLMapFlags`*** | | | +| 0 |*`cudaGLMapFlagsNone`* | | | +| 1 |*`cudaGLMapFlagsReadOnly`* | | | +| 2 |*`cudaGLMapFlagsWriteDiscard`* | | | +| enum |***`cudaD3D9DeviceList`*** | | | +| 1 |*`cudaD3D9DeviceListAll`* | | | +| 2 |*`cudaD3D9DeviceListCurrentFrame`* | | | +| 3 |*`cudaD3D9DeviceListNextFrame`* | | | +| enum |***`cudaD3D9MapFlags`*** | | | +| 0 |*`cudaD3D9MapFlagsNone`* | | | +| 1 |*`cudaD3D9MapFlagsReadOnly`* | | | +| 2 |*`cudaD3D9MapFlagsWriteDiscard`* | | | +| enum |***`cudaD3D9RegisterFlags`*** | | | +| 0 |*`cudaD3D9RegisterFlagsNone`* | | | +| 1 |*`cudaD3D9RegisterFlagsArray`* | | | +| enum |***`cudaD3D10DeviceList`*** | | | +| 1 |*`cudaD3D10DeviceListAll`* | | | +| 2 |*`cudaD3D10DeviceListCurrentFrame`* | | | +| 3 |*`cudaD3D10DeviceListNextFrame`* | | | +| enum |***`cudaD3D10MapFlags`*** | | | +| 0 |*`cudaD3D10MapFlagsNone`* | | | +| 1 |*`cudaD3D10MapFlagsReadOnly`* | | | +| 2 |*`cudaD3D10MapFlagsWriteDiscard`* | | | +| enum |***`cudaD3D10RegisterFlags`*** | | | +| 0 |*`cudaD3D10RegisterFlagsNone`* | | | +| 1 |*`cudaD3D10RegisterFlagsArray`* | | | +| enum |***`cudaD3D11DeviceList`*** | | | +| 1 |*`cudaD3D11DeviceListAll`* | | | +| 2 |*`cudaD3D11DeviceListCurrentFrame`* | | | +| 3 |*`cudaD3D11DeviceListNextFrame`* | | | +| struct |`cudaArray` | |`hipArray` | +| typedef |`cudaArray_t` | |`hipArray_t` | +| typedef |`cudaArray_const_t` | |`hipArray_const_t` | +| typedef |`cudaEvent_t` | |`hipEvent_t` | +| struct |`CUevent_st` | |`ihipEvent_t` | +| typedef |`cudaGraphicsResource_t` | | | +| struct |`cudaMipmappedArray` | |`hipMipmappedArray` | +| typedef |`cudaMipmappedArray_t` | |`hipMipmappedArray_t` | +| typedef |`cudaMipmappedArray_const_t` | |`hipMipmappedArray_const_t` | +| enum |***`cudaOutputMode`*** | | | +| typedef |***`cudaOutputMode_t`*** | | | +| 0x00 |*`cudaKeyValuePair`* | | | +| 0x01 |*`cudaCSV`* | | | +| typedef |`cudaStream_t` | |`hipStream_t` | +| struct |`CUstream_st` | |`ihipStream_t` | +| typedef |`cudaStreamCallback_t` | |`hipStreamCallback_t` | +| typedef |`cudaSurfaceObject_t` | |`hipSurfaceObject_t` | +| typedef |`cudaTextureObject_t` | |`hipTextureObject_t` | +| struct |`CUuuid_st` | | | +| typedef |`cudaUUID_t` | | | +| define |`CUDA_EGL_MAX_PLANES` | 9.1 | | +| define |`CUDA_IPC_HANDLE_SIZE` | | | +| define |`cudaArrayColorAttachment` | 10.0 | | +| define |`cudaArrayCubemap` | |`hipArrayCubemap` | +| define |`cudaArrayDefault` | |`hipArrayDefault` | +| define |`cudaArrayLayered` | |`hipArrayLayered` | +| define |`cudaArraySurfaceLoadStore` | |`hipArraySurfaceLoadStore` | +| define |`cudaArrayTextureGather` | |`hipArrayTextureGather` | +| define |`cudaCooperativeLaunchMultiDeviceNoPreSync` | 9.0 | | +| define |`cudaCooperativeLaunchMultiDeviceNoPostSync` | 9.0 | | +| define |`cudaCpuDeviceId` | 8.0 | | +| define |`cudaInvalidDeviceId` | 8.0 | | +| define |`cudaDeviceBlockingSync` | |`hipDeviceScheduleBlockingSync` | +| define |`cudaDeviceLmemResizeToMax` | |`hipDeviceLmemResizeToMax` | 0x16 | +| define |`cudaDeviceMapHost` | |`hipDeviceMapHost` | +| define |`cudaDeviceMask` | | | +| define |`cudaDevicePropDontCare` | | | +| define |`cudaDeviceScheduleAuto` | |`hipDeviceScheduleAuto` | +| define |`cudaDeviceScheduleBlockingSync` | |`hipDeviceScheduleBlockingSync` | +| define |`cudaDeviceScheduleMask` | |`hipDeviceScheduleMask` | +| define |`cudaDeviceScheduleSpin` | |`hipDeviceScheduleSpin` | +| define |`cudaDeviceScheduleYield` | |`hipDeviceScheduleYield` | +| define |`cudaEventDefault` | |`hipEventDefault` | +| define |`cudaEventBlockingSync` | |`hipEventBlockingSync` | +| define |`cudaEventDisableTiming` | |`hipEventDisableTiming` | +| define |`cudaEventInterprocess` | |`hipEventInterprocess` | +| define |`cudaHostAllocDefault` | |`hipHostMallocDefault` | +| define |`cudaHostAllocMapped` | |`hipHostMallocMapped` | +| define |`cudaHostAllocPortable` | |`hipHostMallocPortable` | +| define |`cudaHostAllocWriteCombined` | |`hipHostMallocWriteCombined` | +| define |`cudaHostRegisterDefault` | |`hipHostRegisterDefault` | +| define |`cudaHostRegisterIoMemory` | 7.5 |`hipHostRegisterIoMemory` | +| define |`cudaHostRegisterMapped` | |`hipHostRegisterMapped` | +| define |`cudaHostRegisterPortable` | |`hipHostRegisterPortable` | +| define |`cudaIpcMemLazyEnablePeerAccess` | |`hipIpcMemLazyEnablePeerAccess` | 0 | +| define |`cudaMemAttachGlobal` | | | +| define |`cudaMemAttachHost` | | | +| define |`cudaMemAttachSingle` | | | +| define |`cudaOccupancyDefault` | | | +| define |`cudaOccupancyDisableCachingOverride` | | | +| define |`cudaPeerAccessDefault` | | | +| define |`cudaStreamDefault` | |`hipStreamDefault` | +| define |`cudaStreamNonBlocking` | |`hipStreamNonBlocking` | +| define |`cudaStreamLegacy` | | | +| define |`cudaStreamPerThread` | | | +| define |`cudaTextureType1D` | |`hipTextureType1D` | +| define |`cudaTextureType2D` | |`hipTextureType2D` | +| define |`cudaTextureType3D` | |`hipTextureType3D` | +| define |`cudaTextureTypeCubemap` | |`hipTextureTypeCubemap` | +| define |`cudaTextureType1DLayered` | |`hipTextureType1DLayered` | +| define |`cudaTextureType2DLayered` | |`hipTextureType2DLayered` | +| define |`cudaTextureTypeCubemapLayered` | |`hipTextureTypeCubemapLayered` | +| enum |***`cudaDataType_t`*** | 8.0 |***`hipblasDatatype_t`*** | +| enum |***`cudaDataType`*** | 8.0 |***`hipblasDatatype_t`*** | +| 2 |*`CUDA_R_16F`* | 8.0 |*`HIPBLAS_R_16F`* | 150 | +| 6 |*`CUDA_C_16F`* | 8.0 |*`HIPBLAS_C_16F`* | 153 | +| 0 |*`CUDA_R_32F`* | 8.0 |*`HIPBLAS_R_32F`* | 151 | +| 4 |*`CUDA_C_32F`* | 8.0 |*`HIPBLAS_C_32F`* | 154 | +| 1 |*`CUDA_R_64F`* | 8.0 |*`HIPBLAS_R_64F`* | 152 | +| 5 |*`CUDA_C_64F`* | 8.0 |*`HIPBLAS_C_64F`* | 155 | +| 3 |*`CUDA_R_8I`* | 8.0 | | +| 7 |*`CUDA_C_8I`* | 8.0 | | +| 8 |*`CUDA_R_8U`* | 8.0 | | +| 9 |*`CUDA_C_8U`* | 8.0 | | +| 10 |*`CUDA_R_32I`* | 8.0 | | +| 11 |*`CUDA_C_32I`* | 8.0 | | +| 12 |*`CUDA_R_32U`* | 8.0 | | +| 13 |*`CUDA_C_32U`* | 8.0 | | +| struct |`cudaExternalMemoryBufferDesc` | 10.0 | | +| struct |`cudaExternalMemoryHandleDesc` | 10.0 | | +| struct |`cudaExternalMemoryMipmappedArrayDesc` | 10.0 | | +| struct |`cudaExternalSemaphoreHandleDesc` | 10.0 | | +| struct |`cudaExternalSemaphoreSignalParams` | 10.0 | | +| struct |`cudaExternalSemaphoreWaitParams` | 10.0 | | +| struct |`cudaHostNodeParams` | 10.0 | | +| struct |`cudaLaunchParams` | 9.0 | | +| struct |`cudaMemsetParams` | 10.0 | | +| struct |`CUeglStreamConnection_st` | 9.1 | | +| typedef |`cudaEglStreamConnection` | 9.1 | | +| define |`cudaExternalMemoryDedicated` | 10.0 | | +| typedef |`cudaExternalMemory_t` | 10.0 | | +| struct |`CUexternalMemory_st` | 10.0 | | +| typedef |`cudaExternalSemaphore_t` | 10.0 | | +| struct |`CUexternalSemaphore_st` | 10.0 | | +| typedef |`cudaGraph_t` | 10.0 | | +| struct |`CUgraph_st` | 10.0 | | +| typedef |`cudaGraphNode_t` | 10.0 | | +| struct |`CUgraphNode_st` | 10.0 | | +| typedef |`cudaGraphExec_t` | 10.0 | | +| struct |`CUgraphExec_st` | 10.0 | | +| typedef |`cudaGraphicsResource_t` | | | +| struct |`cudaGraphicsResource` | | | +| typedef |`cudaHostFn_t` | 10.0 | | +| enum |***`libraryPropertyType`*** | 8.0 | | +| typedef |***`libraryPropertyType_t`*** | 8.0 | | +| 0 |*`MAJOR_VERSION`* | 8.0 | | +| 1 |*`MINOR_VERSION`* | 8.0 | | +| 2 |*`PATCH_LEVEL`* | 8.0 | | + +\* CUDA version, in which API has appeared and (optional) last version before abandoning it; no value in case of earlier versions < 7.5. diff --git a/projects/hip/hipify-clang/src/CUDA2HIP_Runtime_API_types.cpp b/projects/hip/hipify-clang/src/CUDA2HIP_Runtime_API_types.cpp index 662655edb9..e8956b6956 100644 --- a/projects/hip/hipify-clang/src/CUDA2HIP_Runtime_API_types.cpp +++ b/projects/hip/hipify-clang/src/CUDA2HIP_Runtime_API_types.cpp @@ -1067,7 +1067,7 @@ const std::map CUDA_RUNTIME_TYPE_NAME_MAP { // CU_SHAREDMEM_CARVEOUT_MAX_SHARED {"cudaSharedmemCarveoutMaxShared", {"hipSharedmemCarveoutMaxShared", "", CONV_NUMERIC_LITERAL, API_RUNTIME, HIP_UNSUPPORTED}}, // 100 // CU_SHAREDMEM_CARVEOUT_MAX_L1 - {"cudaSharedmemCarveoutMaxShared", {"hipSharedmemCarveoutMaxL1", "", CONV_NUMERIC_LITERAL, API_RUNTIME, HIP_UNSUPPORTED}}, // 0 + {"cudaSharedmemCarveoutMaxL1", {"hipSharedmemCarveoutMaxL1", "", CONV_NUMERIC_LITERAL, API_RUNTIME, HIP_UNSUPPORTED}}, // 0 // CUsharedconfig {"cudaSharedMemConfig", {"hipSharedMemConfig", "", CONV_TYPE, API_RUNTIME}}, @@ -1244,6 +1244,9 @@ const std::map CUDA_RUNTIME_TYPE_NAME_MAP { // CUtexObject {"cudaTextureObject_t", {"hipTextureObject_t", "", CONV_TYPE, API_RUNTIME}}, + // CUuuid + {"cudaUUID_t", {"hipUUID_t", "", CONV_TYPE, API_RUNTIME, HIP_UNSUPPORTED}}, + // 5. Defines // no analogue From 76ea65c212581349da78551921a492781487c39b Mon Sep 17 00:00:00 2001 From: "Yaxun (Sam) Liu" Date: Wed, 31 Jul 2019 16:25:15 -0400 Subject: [PATCH 07/11] Fix -std=c++14 for windows [ROCm/hip commit: f467cb850e463ec5a23ec5c39032c0618c1efc57] --- projects/hip/bin/hipcc | 11 +++++++---- 1 file changed, 7 insertions(+), 4 deletions(-) diff --git a/projects/hip/bin/hipcc b/projects/hip/bin/hipcc index 85e090c761..1eda99b936 100755 --- a/projects/hip/bin/hipcc +++ b/projects/hip/bin/hipcc @@ -195,7 +195,12 @@ if ($HIP_PLATFORM eq "clang") { print ("DEVICE_LIB_PATH=$DEVICE_LIB_PATH\n"); } - $HIPCXXFLAGS .= " -std=c++11 -isystem $HIP_CLANG_INCLUDE_PATH"; + if ($isWindows) { + $HIPCXXFLAGS .= " -std=c++14 -fms-extensions -fms-compatibility"; + } else { + $HIPCXXFLAGS .= " -std=c++11"; + } + $HIPCXXFLAGS .= " -isystem $HIP_CLANG_INCLUDE_PATH"; $HIPLDFLAGS .= " -L$HIP_LIB_PATH"; if (not $isWindows) { $HIPLDFLAGS .= " -Wl,--rpath-link=$HIP_LIB_PATH"; @@ -876,9 +881,7 @@ if ($HIP_PLATFORM eq "clang") { } $HIP_DEVLIB_FLAGS = " --hip-device-lib-path=$DEVICE_LIB_PATH"; $HIPCXXFLAGS .= " $HIP_DEVLIB_FLAGS"; - if ($isWindows) { - $HIPCXXFLAGS .= " -std=c++14 -fms-extensions -fms-compatibility"; - } else { + if (not $isWindows) { $HIPLDFLAGS .= " -lgcc_s -lgcc -lpthread -lm"; } } From 6837312da446bba9ce094badebb011c979f151c0 Mon Sep 17 00:00:00 2001 From: Vladislav Sytchenko Date: Wed, 31 Jul 2019 17:04:31 -0400 Subject: [PATCH 08/11] Don't use a hardcoded warp size, since it can be dynamically changed. Query it from the runtime instead. [ROCm/hip commit: fd3b6263864bcc3e57ba5d72ed6a9d29a2d83042] --- projects/hip/tests/src/deviceLib/hip_mbcnt.cpp | 10 +++++----- 1 file changed, 5 insertions(+), 5 deletions(-) diff --git a/projects/hip/tests/src/deviceLib/hip_mbcnt.cpp b/projects/hip/tests/src/deviceLib/hip_mbcnt.cpp index 343d6069d4..cd4bfa5daa 100644 --- a/projects/hip/tests/src/deviceLib/hip_mbcnt.cpp +++ b/projects/hip/tests/src/deviceLib/hip_mbcnt.cpp @@ -59,12 +59,12 @@ int main() { cout << "hip Device prop succeeded " << endl; - constexpr unsigned int wave_size = 64; constexpr unsigned int num_waves_per_block = 2; - constexpr unsigned int num_threads_per_block = wave_size * num_waves_per_block; - constexpr unsigned int num_blocks = 2; - constexpr unsigned int num_threads = num_threads_per_block * num_blocks; - constexpr size_t buffer_size = num_threads * sizeof(unsigned int); + const unsigned int wave_size = devProp.warpSize; + const unsigned int num_threads_per_block = wave_size * num_waves_per_block; + const unsigned int num_blocks = 2; + const unsigned int num_threads = num_threads_per_block * num_blocks; + const size_t buffer_size = num_threads * sizeof(unsigned int); HIP_ASSERT(hipMalloc((void**)&device_mbcnt_lo, buffer_size)); HIP_ASSERT(hipMalloc((void**)&device_mbcnt_hi, buffer_size)); From 1e92f785bce762f65696d82d59c078325b09d7e5 Mon Sep 17 00:00:00 2001 From: Maneesh Gupta Date: Thu, 1 Aug 2019 13:16:39 +0530 Subject: [PATCH 09/11] [nvcc] Populate missing fields in hipGetDeviceProperties Change-Id: Ie90e02674d503e385f144f1ead3d53ff7b49cecc [ROCm/hip commit: 3ec381d7293fab3346f37b12e45b0f2ca6b7db40] --- .../include/hip/nvcc_detail/hip_runtime_api.h | 25 ++++++++----------- 1 file changed, 11 insertions(+), 14 deletions(-) diff --git a/projects/hip/include/hip/nvcc_detail/hip_runtime_api.h b/projects/hip/include/hip/nvcc_detail/hip_runtime_api.h index 2c294c91a3..fb4987feee 100644 --- a/projects/hip/include/hip/nvcc_detail/hip_runtime_api.h +++ b/projects/hip/include/hip/nvcc_detail/hip_runtime_api.h @@ -715,17 +715,20 @@ inline static hipError_t hipGetDeviceProperties(hipDeviceProp_t* p_prop, int dev struct cudaDeviceProp cdprop; cudaError_t cerror; cerror = cudaGetDeviceProperties(&cdprop, device); + strncpy(p_prop->name, cdprop.name, 256); p_prop->totalGlobalMem = cdprop.totalGlobalMem; p_prop->sharedMemPerBlock = cdprop.sharedMemPerBlock; p_prop->regsPerBlock = cdprop.regsPerBlock; p_prop->warpSize = cdprop.warpSize; + p_prop->maxThreadsPerBlock = cdprop.maxThreadsPerBlock; for (int i = 0; i < 3; i++) { p_prop->maxThreadsDim[i] = cdprop.maxThreadsDim[i]; p_prop->maxGridSize[i] = cdprop.maxGridSize[i]; } - p_prop->maxThreadsPerBlock = cdprop.maxThreadsPerBlock; p_prop->clockRate = cdprop.clockRate; + p_prop->memoryClockRate = cdprop.memoryClockRate; + p_prop->memoryBusWidth = cdprop.memoryBusWidth; p_prop->totalConstMem = cdprop.totalConstMem; p_prop->major = cdprop.major; p_prop->minor = cdprop.minor; @@ -733,44 +736,38 @@ inline static hipError_t hipGetDeviceProperties(hipDeviceProp_t* p_prop, int dev p_prop->l2CacheSize = cdprop.l2CacheSize; p_prop->maxThreadsPerMultiProcessor = cdprop.maxThreadsPerMultiProcessor; p_prop->computeMode = cdprop.computeMode; - p_prop->canMapHostMemory = cdprop.canMapHostMemory; - p_prop->memoryClockRate = cdprop.memoryClockRate; - p_prop->memoryBusWidth = cdprop.memoryBusWidth; - - // Same as clock-rate: - p_prop->clockInstructionRate = cdprop.clockRate; + p_prop->clockInstructionRate = cdprop.clockRate; // Same as clock-rate: int ccVers = p_prop->major * 100 + p_prop->minor * 10; - p_prop->arch.hasGlobalInt32Atomics = (ccVers >= 110); p_prop->arch.hasGlobalFloatAtomicExch = (ccVers >= 110); p_prop->arch.hasSharedInt32Atomics = (ccVers >= 120); p_prop->arch.hasSharedFloatAtomicExch = (ccVers >= 120); - p_prop->arch.hasFloatAtomicAdd = (ccVers >= 200); - p_prop->arch.hasGlobalInt64Atomics = (ccVers >= 120); p_prop->arch.hasSharedInt64Atomics = (ccVers >= 110); - p_prop->arch.hasDoubles = (ccVers >= 130); - p_prop->arch.hasWarpVote = (ccVers >= 120); p_prop->arch.hasWarpBallot = (ccVers >= 200); p_prop->arch.hasWarpShuffle = (ccVers >= 300); p_prop->arch.hasFunnelShift = (ccVers >= 350); - p_prop->arch.hasThreadFenceSystem = (ccVers >= 200); p_prop->arch.hasSyncThreadsExt = (ccVers >= 200); - p_prop->arch.hasSurfaceFuncs = (ccVers >= 200); p_prop->arch.has3dGrid = (ccVers >= 200); p_prop->arch.hasDynamicParallelism = (ccVers >= 350); p_prop->concurrentKernels = cdprop.concurrentKernels; + p_prop->pciDomainID = cdprop.pciDomainID; p_prop->pciBusID = cdprop.pciBusID; p_prop->pciDeviceID = cdprop.pciDeviceID; p_prop->maxSharedMemoryPerMultiProcessor = cdprop.sharedMemPerMultiprocessor; p_prop->isMultiGpuBoard = cdprop.isMultiGpuBoard; + p_prop->canMapHostMemory = cdprop.canMapHostMemory; + p_prop->gcnArch = 0; // Not a GCN arch + p_prop->integrated = cdprop.integrated; + p_prop->cooperativeLaunch = cdprop.cooperativeLaunch; + p_prop->cooperativeMultiDeviceLaunch = cdprop.cooperativeMultiDeviceLaunch; p_prop->maxTexture1D = cdprop.maxTexture1D; p_prop->maxTexture2D[0] = cdprop.maxTexture2D[0]; From 7676b86f12581b3b1752938e53e833ae0fa3edeb Mon Sep 17 00:00:00 2001 From: wkwchau Date: Thu, 1 Aug 2019 04:58:48 -0400 Subject: [PATCH 10/11] =?UTF-8?q?Added=20support=20of=20hipOccupancyMaxAct?= =?UTF-8?q?iveBlocksPerMultiprocessor=20&=20hipOc=E2=80=A6=20(#1240)?= MIME-Version: 1.0 Content-Type: text/plain; charset=UTF-8 Content-Transfer-Encoding: 8bit * Added support of hipOccupancyMaxActiveBlocksPerMultiprocessor & hipOccupancyMaxActiveBlocksPerMultiprocessorWithFlags APIs * Taking into account of SGPR usage to determine the max active blocks in hipOccupancyMaxActiveBlocksPerMultiprocessor() [ROCm/hip commit: 4b18b321f722d07db164c6f4040a41ee93feef16] --- .../hip/hcc_detail/functional_grid_launch.hpp | 14 ++ .../include/hip/hcc_detail/hip_runtime_api.h | 23 +--- projects/hip/src/hip_module.cpp | 123 ++++++++++++++---- ...upancyMaxActiveBlocksPerMultiprocessor.cpp | 78 +++++++++++ 4 files changed, 195 insertions(+), 43 deletions(-) create mode 100644 projects/hip/tests/src/runtimeApi/module/hipOccupancyMaxActiveBlocksPerMultiprocessor.cpp diff --git a/projects/hip/include/hip/hcc_detail/functional_grid_launch.hpp b/projects/hip/include/hip/hcc_detail/functional_grid_launch.hpp index 111a6e2b70..f502fddf42 100644 --- a/projects/hip/include/hip/hcc_detail/functional_grid_launch.hpp +++ b/projects/hip/include/hip/hcc_detail/functional_grid_launch.hpp @@ -151,6 +151,20 @@ void hipOccupancyMaxPotentialBlockSize(uint32_t* gridSize, uint32_t* blockSize, dynSharedMemPerBlk, blockSizeLimit); } +template +inline +void hipOccupancyMaxActiveBlocksPerMultiprocessor(uint32_t* numBlocks, F kernel, + uint32_t blockSize, size_t dynSharedMemPerBlk) { + + using namespace hip_impl; + + hip_impl::hip_init(); + auto f = get_program_state().kernel_descriptor(reinterpret_cast(kernel), + target_agent(0)); + + hipOccupancyMaxActiveBlocksPerMultiprocessor(numBlocks, f, blockSize, dynSharedMemPerBlk); +} + template inline void hipLaunchKernelGGL(F kernel, const dim3& numBlocks, const dim3& dimBlocks, diff --git a/projects/hip/include/hip/hcc_detail/hip_runtime_api.h b/projects/hip/include/hip/hcc_detail/hip_runtime_api.h index 979efebe4d..7e887d1172 100644 --- a/projects/hip/include/hip/hcc_detail/hip_runtime_api.h +++ b/projects/hip/include/hip/hcc_detail/hip_runtime_api.h @@ -2749,7 +2749,7 @@ hipError_t hipLaunchCooperativeKernelMultiDevice(hipLaunchParams* launchParamsLi * * @param [out] gridSize minimum grid size for maximum potential occupancy * @param [out] blockSize block size for maximum potential occupancy - * @param [in] f kernel to launch + * @param [in] f kernel function for which occupancy is calulated * @param [in] dynSharedMemPerBlk dynamic shared memory usage (in bytes) intended for each block * @param [in] blockSizeLimit the maximum block size for the kernel, use 0 for no limit * @@ -2765,10 +2765,10 @@ hipError_t hipOccupancyMaxPotentialBlockSize(uint32_t* gridSize, uint32_t* block * @param [out] numBlocks Returned occupancy * @param [in] func Kernel function for which occupancy is calulated * @param [in] blockSize Block size the kernel is intended to be launched with - * @param [in] dynamicSMemSize Per - block dynamic shared memory usage intended, in bytes + * @param [in] dynSharedMemPerBlk dynamic shared memory usage (in bytes) intended for each block */ hipError_t hipOccupancyMaxActiveBlocksPerMultiprocessor( - int* numBlocks, const void* f, int blockSize, size_t dynamicSMemSize); + uint32_t* numBlocks, hipFunction_t f, uint32_t blockSize, size_t dynSharedMemPerBlk); /** * @brief Returns occupancy for a device function. @@ -2776,11 +2776,11 @@ hipError_t hipOccupancyMaxActiveBlocksPerMultiprocessor( * @param [out] numBlocks Returned occupancy * @param [in] func Kernel function for which occupancy is calulated * @param [in] blockSize Block size the kernel is intended to be launched with - * @param [in] dynamicSMemSize Per - block dynamic shared memory usage intended, in bytes + * @param [in] dynSharedMemPerBlk dynamic shared memory usage (in bytes) intended for each block * @param [in] flags Extra flags for occupancy calculation (currently ignored) */ hipError_t hipOccupancyMaxActiveBlocksPerMultiprocessorWithFlags( - int* numBlocks, const void* f, int blockSize, size_t dynamicSMemSize, unsigned int flags); + uint32_t* numBlocks, hipFunction_t f, uint32_t blockSize, size_t dynSharedMemPerBlk, unsigned int flags); /** * @brief Launches kernels on multiple devices and guarantees all specified kernels are dispatched @@ -3123,19 +3123,6 @@ hipError_t hipBindTextureToMipmappedArray(const texture& tex, return hipSuccess; } -template -inline hipError_t hipOccupancyMaxActiveBlocksPerMultiprocessor( - int* numBlocks, T f, int blockSize, size_t dynamicSMemSize) { - return hipOccupancyMaxActiveBlocksPerMultiprocessor( - numBlocks, reinterpret_cast(f), blockSize, dynamicSMemSize); -} - -template -inline hipError_t hipOccupancyMaxActiveBlocksPerMultiprocessorWithFlags( - int* numBlocks, T f, int blockSize, size_t dynamicSMemSize, unsigned int flags) { - return hipOccupancyMaxActiveBlocksPerMultiprocessorWithFlags( - numBlocks, reinterpret_cast(f), blockSize, dynamicSMemSize, flags); -} template inline hipError_t hipLaunchCooperativeKernel(T f, dim3 gridDim, dim3 blockDim, diff --git a/projects/hip/src/hip_module.cpp b/projects/hip/src/hip_module.cpp index 2d3d052036..9e972da246 100644 --- a/projects/hip/src/hip_module.cpp +++ b/projects/hip/src/hip_module.cpp @@ -879,6 +879,30 @@ hipError_t hipModuleGetTexRef(textureReference** texRef, hipModule_t hmod, const return ihipLogStatus(hipSuccess); } +void getGprsLdsUsage(hipFunction_t f, size_t* usedVGPRS, size_t* usedSGPRS, size_t* usedLDS) +{ + bool is_code_object_v3 = f->_name.find(".kd") != std::string::npos; + if (is_code_object_v3) { + const auto header = reinterpret_cast(f->_header); + // GRANULATED_WAVEFRONT_VGPR_COUNT is specified in 0:5 bits of COMPUTE_PGM_RSRC1 + // the granularity for gfx6-gfx9 is max(0, ceil(vgprs_used / 4) - 1) + *usedVGPRS = ((header->compute_pgm_rsrc1 & 0x3F) + 1) << 2; + // GRANULATED_WAVEFRONT_SGPR_COUNT is specified in 6:9 bits of COMPUTE_PGM_RSRC1 + // the granularity for gfx9+ is 2 * max(0, ceil(sgprs_used / 16) - 1) + *usedSGPRS = ((((header->compute_pgm_rsrc1 & 0x3C0) >> 6) >> 1) + 1) << 4; + *usedLDS = header->group_segment_fixed_size; + } + else { + const auto header = f->_header; + // VGPRs granularity is 4 + *usedVGPRS = ((header->workitem_vgpr_count + 3) >> 2) << 2; + // adding 2 to take into account the 2 VCC registers & handle the granularity of 16 + *usedSGPRS = header->wavefront_sgpr_count + 2; + *usedSGPRS = ((*usedSGPRS + 15) >> 4) << 4; + *usedLDS = header->workgroup_group_segment_byte_size; + } +} + hipError_t ihipOccupancyMaxPotentialBlockSize(uint32_t* gridSize, uint32_t* blockSize, hipFunction_t f, size_t dynSharedMemPerBlk, uint32_t blockSizeLimit) @@ -886,10 +910,8 @@ hipError_t ihipOccupancyMaxPotentialBlockSize(uint32_t* gridSize, uint32_t* bloc using namespace hip_impl; auto ctx = ihipGetTlsDefaultCtx(); - hipError_t ret = hipSuccess; - if (ctx == nullptr) { - ret = hipErrorInvalidDevice; + return hipErrorInvalidDevice; } hipDeviceProp_t prop{}; @@ -900,26 +922,7 @@ hipError_t ihipOccupancyMaxPotentialBlockSize(uint32_t* gridSize, uint32_t* bloc size_t usedVGPRS = 0; size_t usedSGPRS = 0; size_t usedLDS = 0; - bool is_code_object_v3 = f->_name.find(".kd") != std::string::npos; - if (is_code_object_v3) { - const auto header = reinterpret_cast(f->_header); - // GRANULATED_WAVEFRONT_VGPR_COUNT is specified in 0:5 bits of COMPUTE_PGM_RSRC1 - // the granularity for gfx6-gfx9 is max(0, ceil(vgprs_used / 4) - 1) - usedVGPRS = ((header->compute_pgm_rsrc1 & 0x3F) + 1) << 2; - // GRANULATED_WAVEFRONT_SGPR_COUNT is specified in 6:9 bits of COMPUTE_PGM_RSRC1 - // the granularity for gfx9+ is 2 * max(0, ceil(sgprs_used / 16) - 1) - usedSGPRS = ((((header->compute_pgm_rsrc1 & 0x3C0) >> 6) >> 1) + 1) << 4; - usedLDS = header->group_segment_fixed_size; - } - else { - const auto header = f->_header; - // VGPRs granularity is 4 - usedVGPRS = ((header->workitem_vgpr_count + 3) >> 2) << 2; - // adding 2 to take into account the 2 VCC registers & handle the granularity of 16 - usedSGPRS = header->wavefront_sgpr_count + 2; - usedSGPRS = ((usedSGPRS + 15) >> 4) << 4; - usedLDS = header->workgroup_group_segment_byte_size; - } + getGprsLdsUsage(f, &usedVGPRS, &usedSGPRS, &usedLDS); // try different workgroup sizes to find the maximum potential occupancy // based on the usage of VGPRs and LDS @@ -1009,10 +1012,9 @@ hipError_t ihipOccupancyMaxPotentialBlockSize(uint32_t* gridSize, uint32_t* bloc *blockSize = maxWavefronts * wavefrontSize; *gridSize = min((maxThreadsCnt + *blockSize - 1) / *blockSize, prop.multiProcessorCount); - return ret; + return hipSuccess; } - hipError_t hipOccupancyMaxPotentialBlockSize(uint32_t* gridSize, uint32_t* blockSize, hipFunction_t f, size_t dynSharedMemPerBlk, uint32_t blockSizeLimit) @@ -1022,3 +1024,74 @@ hipError_t hipOccupancyMaxPotentialBlockSize(uint32_t* gridSize, uint32_t* block return ihipLogStatus(ihipOccupancyMaxPotentialBlockSize( gridSize, blockSize, f, dynSharedMemPerBlk, blockSizeLimit)); } + +hipError_t ihipOccupancyMaxActiveBlocksPerMultiprocessor( + uint32_t* numBlocks, hipFunction_t f, uint32_t blockSize, size_t dynSharedMemPerBlk) +{ + using namespace hip_impl; + + auto ctx = ihipGetTlsDefaultCtx(); + if (ctx == nullptr) { + return hipErrorInvalidDevice; + } + + hipDeviceProp_t prop{}; + ihipGetDeviceProperties(&prop, ihipGetTlsDefaultCtx()->getDevice()->_deviceId); + + prop.regsPerBlock = prop.regsPerBlock ? prop.regsPerBlock : 64 * 1024; + + size_t usedVGPRS = 0; + size_t usedSGPRS = 0; + size_t usedLDS = 0; + getGprsLdsUsage(f, &usedVGPRS, &usedSGPRS, &usedLDS); + + // Due to SPI and private memory limitations, the max of wavefronts per CU in 32 + size_t wavefrontSize = prop.warpSize; + size_t maxWavefrontsPerCU = min(prop.maxThreadsPerMultiProcessor / wavefrontSize, 32); + + const size_t simdPerCU = 4; + const size_t maxWavesPerSimd = maxWavefrontsPerCU / simdPerCU; + + size_t numWavefronts = (blockSize + wavefrontSize - 1) / wavefrontSize; + + size_t availableVGPRs = (prop.regsPerBlock / wavefrontSize / simdPerCU); + size_t vgprs_alu_occupancy = simdPerCU * std::min(maxWavesPerSimd, availableVGPRs / usedVGPRS); + + // Calculate blocks occupancy per CU based on VGPR usage + *numBlocks = vgprs_alu_occupancy / numWavefronts; + + const size_t availableSGPRs = (prop.gcnArch < 800) ? 512 : 800; + size_t sgprs_alu_occupancy = simdPerCU * ((usedSGPRS == 0) ? maxWavesPerSimd + : std::min(maxWavesPerSimd, availableSGPRs / usedSGPRS)); + + // Calculate blocks occupancy per CU based on SGPR usage + *numBlocks = std::min(*numBlocks, (uint32_t) (sgprs_alu_occupancy / numWavefronts)); + + size_t total_used_lds = usedLDS + dynSharedMemPerBlk; + if (total_used_lds != 0) { + // Calculate LDS occupacy per CU. lds_per_cu / (static_lsd + dynamic_lds) + size_t lds_occupancy = prop.maxSharedMemoryPerMultiProcessor / total_used_lds; + *numBlocks = std::min(*numBlocks, (uint32_t) lds_occupancy); + } + + return hipSuccess; +} + +hipError_t hipOccupancyMaxActiveBlocksPerMultiprocessor( + uint32_t* numBlocks, hipFunction_t f, uint32_t blockSize, size_t dynSharedMemPerBlk) +{ + HIP_INIT_API(hipOccupancyMaxActiveBlocksPerMultiprocessor, numBlocks, f, blockSize, dynSharedMemPerBlk); + + return ihipLogStatus(ihipOccupancyMaxActiveBlocksPerMultiprocessor( + numBlocks, f, blockSize, dynSharedMemPerBlk)); +} + +hipError_t hipOccupancyMaxActiveBlocksPerMultiprocessorWithFlags( + uint32_t* numBlocks, hipFunction_t f, uint32_t blockSize, size_t dynSharedMemPerBlk, + unsigned int flags) +{ + HIP_INIT_API(hipOccupancyMaxActiveBlocksPerMultiprocessorWithFlags, numBlocks, f, blockSize, dynSharedMemPerBlk, flags); + + return ihipLogStatus(ihipOccupancyMaxActiveBlocksPerMultiprocessor( + numBlocks, f, blockSize, dynSharedMemPerBlk)); +} diff --git a/projects/hip/tests/src/runtimeApi/module/hipOccupancyMaxActiveBlocksPerMultiprocessor.cpp b/projects/hip/tests/src/runtimeApi/module/hipOccupancyMaxActiveBlocksPerMultiprocessor.cpp new file mode 100644 index 0000000000..ebf656b72f --- /dev/null +++ b/projects/hip/tests/src/runtimeApi/module/hipOccupancyMaxActiveBlocksPerMultiprocessor.cpp @@ -0,0 +1,78 @@ +/* +Copyright (c) 2019 Advanced Micro Devices, Inc. All rights reserved. + +Permission is hereby granted, free of charge, to any person obtaining a copy +of this software and associated documentation files (the "Software"), to deal +in the Software without restriction, including without limitation the rights +to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +copies of the Software, and to permit persons to whom the Software is +furnished to do so, subject to the following conditions: + +The above copyright notice and this permission notice shall be included in +all copies or substantial portions of the Software. + +THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN +THE SOFTWARE. +*/ +// Test the Grid_Launch syntax. + +/* HIT_START + * BUILD: %t %s ../../test_common.cpp EXCLUDE_HIP_PLATFORM nvcc + * TEST: %t + * HIT_END + */ + +#include "hip/hip_runtime.h" +#include "test_common.h" + +#define fileName "vcpy_kernel.code" +#define kernel_name "hello_world" + + +__global__ void f1(float *a) { *a = 1.0; } + +template +__global__ void f2(T *a) { *a = 1; } + + + +int main(int argc, char* argv[]) { + + // test case for using kernel function pointer + uint32_t gridSize = 0; + uint32_t blockSize = 0; + hipOccupancyMaxPotentialBlockSize(&gridSize, &blockSize, f1, 0, 0); + assert(gridSize != 0 && blockSize != 0); + + uint32_t numBlock = 0; + hipOccupancyMaxActiveBlocksPerMultiprocessor(&numBlock, f1, blockSize, 0); + assert(numBlock != 0); + + + // test case for using kernel function pointer with template + gridSize = 0; + blockSize = 0; + hipOccupancyMaxPotentialBlockSize(&gridSize, &blockSize, f2, 0, 0); + assert(gridSize != 0 && blockSize != 0); + + numBlock = 0; + hipOccupancyMaxActiveBlocksPerMultiprocessor(&numBlock, f2, blockSize, 0); + assert(numBlock != 0); + + + // test case for using kernel with hipFunction_t type + numBlock = 0; + hipModule_t Module; + hipFunction_t Function; + HIPCHECK(hipModuleLoad(&Module, fileName)); + HIPCHECK(hipModuleGetFunction(&Function, Module, kernel_name)); + HIPCHECK(hipOccupancyMaxActiveBlocksPerMultiprocessor(&numBlock, Function, blockSize, 0)); + assert(numBlock != 0); + + passed(); +} From c666fdaa086d9be2add2f66244ce22bedccb5a56 Mon Sep 17 00:00:00 2001 From: wkwchau Date: Thu, 1 Aug 2019 12:03:35 -0400 Subject: [PATCH 11/11] =?UTF-8?q?Added=20query=20of=20hipDeviceAttributeHd?= =?UTF-8?q?pMemFlushCntl=20and=20hipDeviceAttribu=E2=80=A6=20(#1238)?= MIME-Version: 1.0 Content-Type: text/plain; charset=UTF-8 Content-Transfer-Encoding: 8bit * Added query of hipDeviceAttributeHdpMemFlushCntl and hipDeviceAttributeHdpRegFlushCntl * Added NVCC blocker for the hip*FlushCntl test cases [ROCm/hip commit: e7447d5809cc50f5660ae704f5ba90af83e06c1f] --- projects/hip/include/hip/hip_runtime_api.h | 8 ++++++-- projects/hip/src/hip_device.cpp | 11 ++++++++++ projects/hip/src/hip_hcc.cpp | 9 +++++++++ .../device/hipGetDeviceAttribute.cpp | 20 +++++++++++++++++++ 4 files changed, 46 insertions(+), 2 deletions(-) diff --git a/projects/hip/include/hip/hip_runtime_api.h b/projects/hip/include/hip/hip_runtime_api.h index 637de712b3..c067f5bf54 100644 --- a/projects/hip/include/hip/hip_runtime_api.h +++ b/projects/hip/include/hip/hip_runtime_api.h @@ -117,10 +117,12 @@ typedef struct hipDeviceProp_t { int integrated; ///< APU vs dGPU int cooperativeLaunch; ///< HIP device supports cooperative launch int cooperativeMultiDeviceLaunch; ///< HIP device supports cooperative launch on multiple devices -#if !__HIP_VDI__ // Temporarily disable the following three new fields for HIP/VDI runtime +#if !__HIP_VDI__ // Temporarily disable the following five new fields for HIP/VDI runtime int maxTexture1D; ///< Maximum number of elements in 1D images int maxTexture2D[2]; ///< Maximum dimensions (width, height) of 2D images, in image elements int maxTexture3D[3]; ///< Maximum dimensions (width, height, depth) of 3D images, in image elements + unsigned int* hdpMemFlushCntl; ///< Addres of HDP_MEM_COHERENCY_FLUSH_CNTL register + unsigned int* hdpRegFlushCntl; ///< Addres of HDP_REG_COHERENCY_FLUSH_CNTL register #endif } hipDeviceProp_t; @@ -306,8 +308,10 @@ typedef enum hipDeviceAttribute_t { hipDeviceAttributeMaxTexture2DHeight, ///< Maximum dimension height of 2D images in image elements hipDeviceAttributeMaxTexture3DWidth, ///< Maximum dimension width of 3D images in image elements hipDeviceAttributeMaxTexture3DHeight, ///< Maximum dimensions height of 3D images in image elements - hipDeviceAttributeMaxTexture3DDepth ///< Maximum dimensions depth of 3D images in image elements + hipDeviceAttributeMaxTexture3DDepth, ///< Maximum dimensions depth of 3D images in image elements + hipDeviceAttributeHdpMemFlushCntl, ///< Address of the HDP_MEM_COHERENCY_FLUSH_CNTL register + hipDeviceAttributeHdpRegFlushCntl ///< Address of the HDP_REG_COHERENCY_FLUSH_CNTL register } hipDeviceAttribute_t; enum hipComputeMode { diff --git a/projects/hip/src/hip_device.cpp b/projects/hip/src/hip_device.cpp index 5c19ad42c6..1c0e5c9109 100644 --- a/projects/hip/src/hip_device.cpp +++ b/projects/hip/src/hip_device.cpp @@ -293,6 +293,17 @@ hipError_t ihipDeviceGetAttribute(int* pi, hipDeviceAttribute_t attr, int device break; case hipDeviceAttributeMaxTexture3DDepth: *pi = prop->maxTexture3D[2]; + case hipDeviceAttributeHdpMemFlushCntl: + { + uint32_t** hdp = reinterpret_cast(pi); + *hdp = prop->hdpMemFlushCntl; + } + break; + case hipDeviceAttributeHdpRegFlushCntl: + { + uint32_t** hdp = reinterpret_cast(pi); + *hdp = prop->hdpRegFlushCntl; + } break; default: e = hipErrorInvalidValue; diff --git a/projects/hip/src/hip_hcc.cpp b/projects/hip/src/hip_hcc.cpp index 1b2f895f7d..f3d6eae6b7 100644 --- a/projects/hip/src/hip_hcc.cpp +++ b/projects/hip/src/hip_hcc.cpp @@ -915,6 +915,15 @@ hipError_t ihipDevice_t::initProperties(hipDeviceProp_t* prop) { err = hsa_agent_get_info(_hsaAgent, (hsa_agent_info_t)HSA_EXT_AGENT_INFO_IMAGE_3D_MAX_ELEMENTS, prop->maxTexture3D); DeviceErrorCheck(err); + + // Get Agent HDP Flush Register Memory + hsa_amd_hdp_flush_t hdpinfo; + err = hsa_agent_get_info(_hsaAgent, (hsa_agent_info_t)HSA_AMD_AGENT_INFO_HDP_FLUSH, &hdpinfo); + DeviceErrorCheck(err); + + prop->hdpMemFlushCntl = hdpinfo.HDP_MEM_FLUSH_CNTL; + prop->hdpRegFlushCntl = hdpinfo.HDP_REG_FLUSH_CNTL; + return e; } diff --git a/projects/hip/tests/src/runtimeApi/device/hipGetDeviceAttribute.cpp b/projects/hip/tests/src/runtimeApi/device/hipGetDeviceAttribute.cpp index 1e1076b806..27947339cb 100644 --- a/projects/hip/tests/src/runtimeApi/device/hipGetDeviceAttribute.cpp +++ b/projects/hip/tests/src/runtimeApi/device/hipGetDeviceAttribute.cpp @@ -56,6 +56,22 @@ hipError_t test_hipDeviceGetAttribute(int deviceId, hipDeviceAttribute_t attr, return hipSuccess; } +hipError_t test_hipDeviceGetHdpAddress(int deviceId, hipDeviceAttribute_t attr, + uint32_t* expectedValue = (uint32_t*)0xdeadbeef) { + uint32_t* value = 0; + std::cout << "Test hipDeviceGetHdpAddress attribute " << attr; + if (expectedValue != (uint32_t*)0xdeadbeef) { + std::cout << " expected value " << expectedValue; + } + hipError_t e = hipDeviceGetAttribute((int*) &value, attr, deviceId); + std::cout << " actual value " << value << std::endl; + if ((expectedValue != (uint32_t*)0xdeadbeef) && value != expectedValue) { + std::cout << "fail" << std::endl; + return hipErrorInvalidValue; + } + return hipSuccess; +} + int main(int argc, char* argv[]) { int deviceId; CHECK(hipGetDevice(&deviceId)); @@ -116,5 +132,9 @@ int main(int argc, char* argv[]) { CHECK(test_hipDeviceGetAttribute(deviceId, hipDeviceAttributeMaxTexture3DWidth, props.maxTexture3D[0])); CHECK(test_hipDeviceGetAttribute(deviceId, hipDeviceAttributeMaxTexture3DHeight, props.maxTexture3D[1])); CHECK(test_hipDeviceGetAttribute(deviceId, hipDeviceAttributeMaxTexture3DDepth, props.maxTexture3D[2])); +#ifndef __HIP_PLATFORM_NVCC__ + CHECK(test_hipDeviceGetHdpAddress(deviceId, hipDeviceAttributeHdpMemFlushCntl, props.hdpMemFlushCntl)); + CHECK(test_hipDeviceGetHdpAddress(deviceId, hipDeviceAttributeHdpRegFlushCntl, props.hdpRegFlushCntl)); +#endif passed(); };