diff --git a/plugins/cublas/backend/CublasHandler.h b/plugins/cublas/backend/CublasHandler.h
index 4547e48f..bf6cc3fa 100644
--- a/plugins/cublas/backend/CublasHandler.h
+++ b/plugins/cublas/backend/CublasHandler.h
@@ -40,7 +40,8 @@
 #include <string>
 #include <memory>
 
-#include <cublas.h>
+// Deprecated to include both in 2.1.15.cuBLAS: Release 12.0
+// #include <cublas.h>
 #include "cublas_v2.h"
 
 #include <gvirtus/backend/Handler.h>
diff --git a/plugins/cublas/backend/CublasHandler_Level2.cpp b/plugins/cublas/backend/CublasHandler_Level2.cpp
index 9a24dc06..2ed855b8 100644
--- a/plugins/cublas/backend/CublasHandler_Level2.cpp
+++ b/plugins/cublas/backend/CublasHandler_Level2.cpp
@@ -852,7 +852,7 @@ CUBLAS_ROUTINE_HANDLER(Ssbmv_v2){
     float * y = in->GetFromMarshal<float*>();
     int incy = in->Get<int>();
     
-    cublasStatus cs = cublasSsbmv_v2(handle,uplo,n,k,alpha,A,lda,x,incx,beta,y,incy);
+    cublasStatus_t cs = cublasSsbmv_v2(handle,uplo,n,k,alpha,A,lda,x,incx,beta,y,incy);
     return std::make_shared<Result>(cs);
 }
 
@@ -874,7 +874,7 @@ CUBLAS_ROUTINE_HANDLER(Dsbmv_v2){
     double * y = in->GetFromMarshal<double*>();
     int incy = in->Get<int>();
     
-    cublasStatus cs = cublasDsbmv_v2(handle,uplo,n,k,alpha,A,lda,x,incx,beta,y,incy);
+    cublasStatus_t cs = cublasDsbmv_v2(handle,uplo,n,k,alpha,A,lda,x,incx,beta,y,incy);
     return std::make_shared<Result>(cs);
 }
 
@@ -896,7 +896,7 @@ CUBLAS_ROUTINE_HANDLER(Chbmv_v2){
     cuComplex * y = in->GetFromMarshal<cuComplex*>();
     int incy = in->Get<int>();
     
-    cublasStatus cs = cublasChbmv_v2(handle,uplo,n,k,alpha,A,lda,x,incx,beta,y,incy);
+    cublasStatus_t cs = cublasChbmv_v2(handle,uplo,n,k,alpha,A,lda,x,incx,beta,y,incy);
     return std::make_shared<Result>(cs);
 }
 
@@ -918,7 +918,7 @@ CUBLAS_ROUTINE_HANDLER(Zhbmv_v2){
     cuDoubleComplex * y = in->GetFromMarshal<cuDoubleComplex*>();
     int incy = in->Get<int>();
     
-    cublasStatus cs = cublasZhbmv_v2(handle,uplo,n,k,alpha,A,lda,x,incx,beta,y,incy);
+    cublasStatus_t cs = cublasZhbmv_v2(handle,uplo,n,k,alpha,A,lda,x,incx,beta,y,incy);
     return std::make_shared<Result>(cs);
 }
 
@@ -938,7 +938,7 @@ CUBLAS_ROUTINE_HANDLER(Sspmv_v2){
     float * y = in->GetFromMarshal<float*>();
     int incy = in->Get<int>();
     
-    cublasStatus cs = cublasSspmv_v2(handle,uplo,n,alpha,AP,x,incx,beta,y,incy);
+    cublasStatus_t cs = cublasSspmv_v2(handle,uplo,n,alpha,AP,x,incx,beta,y,incy);
     return std::make_shared<Result>(cs);
 }
 
@@ -958,7 +958,7 @@ CUBLAS_ROUTINE_HANDLER(Dspmv_v2){
     double * y = in->GetFromMarshal<double*>();
     int incy = in->Get<int>();
     
-    cublasStatus cs = cublasDspmv_v2(handle,uplo,n,alpha,AP,x,incx,beta,y,incy);
+    cublasStatus_t cs = cublasDspmv_v2(handle,uplo,n,alpha,AP,x,incx,beta,y,incy);
     return std::make_shared<Result>(cs);
 }
 
@@ -978,7 +978,7 @@ CUBLAS_ROUTINE_HANDLER(Chpmv_v2){
     cuComplex * y = in->GetFromMarshal<cuComplex*>();
     int incy = in->Get<int>();
     
-    cublasStatus cs = cublasChpmv_v2(handle,uplo,n,alpha,AP,x,incx,beta,y,incy);
+    cublasStatus_t cs = cublasChpmv_v2(handle,uplo,n,alpha,AP,x,incx,beta,y,incy);
     return std::make_shared<Result>(cs);
 }
 
@@ -998,7 +998,7 @@ CUBLAS_ROUTINE_HANDLER(Zhpmv_v2){
     cuDoubleComplex * y = in->GetFromMarshal<cuDoubleComplex*>();
     int incy = in->Get<int>();
     
-    cublasStatus cs = cublasZhpmv_v2(handle,uplo,n,alpha,AP,x,incx,beta,y,incy);
+    cublasStatus_t cs = cublasZhpmv_v2(handle,uplo,n,alpha,AP,x,incx,beta,y,incy);
     return std::make_shared<Result>(cs);
 }
 
diff --git a/plugins/cublas/frontend/CublasFrontend.h b/plugins/cublas/frontend/CublasFrontend.h
index 1908bdc6..2efe69e2 100644
--- a/plugins/cublas/frontend/CublasFrontend.h
+++ b/plugins/cublas/frontend/CublasFrontend.h
@@ -25,8 +25,9 @@
 #ifndef CUBLASFRONTEND_H
 #define	CUBLASFRONTEND_H
 
-#include <cublas.h>
-#include <cublas_api.h>
+// Deprecated to include both in 2.1.15.cuBLAS: Release 12.0
+// #include <cublas.h>
+// #include <cublas_api.h>
 #include "cublas_v2.h"
 #include <cuda_runtime_api.h>
 
diff --git a/plugins/cudart/CMakeLists.txt b/plugins/cudart/CMakeLists.txt
index 914afba1..4e8ab979 100644
--- a/plugins/cudart/CMakeLists.txt
+++ b/plugins/cudart/CMakeLists.txt
@@ -30,6 +30,8 @@ string(REGEX REPLACE ".*\\.so\\.([0-9]+\\.[0-9]+\\.[0-9]+)" "\\1" CUDART_VERSION
 
 message(STATUS "Found cudaRt: ${CUDART_VERSION} (${CUDART_INCLUDE_DIRECTORY}/cuda_runtime_api.h, ${CUDART_LIBRARY})")
 
+if (CUDART_VERSION LESS 12000)
+
 gvirtus_add_backend(cudart ${CUDA_VERSION}
         backend/CudaRtHandler_texture.cpp
         backend/CudaRtHandler_event.cpp
@@ -66,4 +68,44 @@ gvirtus_add_frontend(cudart ${CUDA_VERSION}
         frontend/CudaRt_version.cpp
         util/CudaUtil.cpp)
 
-add_subdirectory(demo)
\ No newline at end of file
+else (CUDART_VERSION LESS 12000)
+
+gvirtus_add_backend(cudart ${CUDA_VERSION}
+        # backend/CudaRtHandler_texture.cpp
+        backend/CudaRtHandler_event.cpp
+        backend/CudaRtHandler_opengl.cpp
+        # backend/CudaRtHandler_surface.cpp
+        backend/CudaRtHandler_thread.cpp
+        backend/CudaRtHandler_internal.cpp
+        backend/CudaRtHandler_occupancy.cpp
+        backend/CudaRtHandler_execution.cpp
+        backend/CudaRtHandler_device.cpp
+        backend/CudaRtHandler_stream.cpp
+        backend/CudaRtHandler_memory.cpp
+        backend/CudaRtHandler_version.cpp
+        backend/CudaRtHandler_error.cpp
+        backend/CudaRtHandler.cpp
+        util/CudaUtil.cpp)
+target_link_libraries(${PROJECT_NAME} ${CUDA_CUDART_LIBRARY})
+
+gvirtus_add_frontend(cudart ${CUDA_VERSION}
+        frontend/CudaRt.cpp
+        frontend/CudaRt_device.cpp
+        frontend/CudaRt_error.cpp
+        frontend/CudaRt_event.cpp
+        frontend/CudaRt_execution.cpp
+        frontend/CudaRtFrontend.cpp
+        frontend/CudaRt_internal.cpp
+        frontend/CudaRt_memory.cpp
+        frontend/CudaRt_occupancy.cpp
+        frontend/CudaRt_opengl.cpp
+        frontend/CudaRt_stream.cpp
+        # frontend/CudaRt_surface.cpp
+        # frontend/CudaRt_texture.cpp
+        frontend/CudaRt_thread.cpp
+        frontend/CudaRt_version.cpp
+        util/CudaUtil.cpp)
+
+endif (CUDART_VERSION LESS 12000)
+
+add_subdirectory(demo)
diff --git a/plugins/cudart/backend/CudaRtHandler.cpp b/plugins/cudart/backend/CudaRtHandler.cpp
index 351bff19..e76ff83c 100644
--- a/plugins/cudart/backend/CudaRtHandler.cpp
+++ b/plugins/cudart/backend/CudaRtHandler.cpp
@@ -59,8 +59,10 @@ CudaRtHandler::CudaRtHandler() {
   mpFatBinary = new map<string, void **>();
   mpDeviceFunction = new map<string, string>();
   mpVar = new map<string, string>();
+#if CUDART_VERSION < 12000
   mpTexture = new map<string, textureReference *>();
   mpSurface = new map<string, surfaceReference *>();
+#endif
 
   mapHost2DeviceFunc = new map<const void*, std::string>();
   mapDeviceFunc2InfoFunc = new map<std::string, NvInfoFunction>();
@@ -208,6 +210,7 @@ const char *CudaRtHandler::GetVar(const char *handler) {
   return GetVar(tmp);
 }
 
+#if CUDART_VERSION < 12000
 void CudaRtHandler::RegisterTexture(string &handler, textureReference *texref) {
   mpTexture->insert(make_pair(handler, texref));
   //#ifdef DEBUG
@@ -276,6 +279,7 @@ const char *CudaRtHandler::GetSurfaceHandler(surfaceReference *surfref) {
     if (it->second == surfref) return it->first.c_str();
   return NULL;
 }
+#endif
 
 const char *CudaRtHandler::GetSymbol(std::shared_ptr<Buffer> in) {
   char *symbol_handler = in->AssignString();
@@ -354,7 +358,7 @@ void CudaRtHandler::Initialize() {
   mspHandlers->insert(CUDA_ROUTINE_HANDLER_PAIR(SetDoubleForDevice));
   mspHandlers->insert(CUDA_ROUTINE_HANDLER_PAIR(SetDoubleForHost));
   mspHandlers->insert(CUDA_ROUTINE_HANDLER_PAIR(SetupArgument));
-#if CUDART_VERSION >= 9020
+#if (CUDART_VERSION >= 9020) && (CUDART_VERSION < 12000)
   mspHandlers->insert(CUDA_ROUTINE_HANDLER_PAIR(PushCallConfiguration));
   mspHandlers->insert(CUDA_ROUTINE_HANDLER_PAIR(PopCallConfiguration));
 #endif
@@ -368,8 +372,10 @@ void CudaRtHandler::Initialize() {
   mspHandlers->insert(CUDA_ROUTINE_HANDLER_PAIR(RegisterVar));
   mspHandlers->insert(CUDA_ROUTINE_HANDLER_PAIR(RegisterSharedVar));
   mspHandlers->insert(CUDA_ROUTINE_HANDLER_PAIR(RegisterShared));
+#if CUDART_VERSION < 12000
   mspHandlers->insert(CUDA_ROUTINE_HANDLER_PAIR(RegisterTexture));
   mspHandlers->insert(CUDA_ROUTINE_HANDLER_PAIR(RegisterSurface));
+#endif
 
   /* CudaRtHandler_memory */
   mspHandlers->insert(CUDA_ROUTINE_HANDLER_PAIR(Free));
@@ -415,31 +421,33 @@ void CudaRtHandler::Initialize() {
   mspHandlers->insert(CUDA_ROUTINE_HANDLER_PAIR(StreamWaitEvent));
   mspHandlers->insert(CUDA_ROUTINE_HANDLER_PAIR(StreamCreateWithPriority));
 
+#ifndef CUDART_VERSION
+#error CUDART_VERSION not defined
+#endif
+#if CUDART_VERSION < 12000
   /* CudaRtHandler_surface */
   mspHandlers->insert(CUDA_ROUTINE_HANDLER_PAIR(BindSurfaceToArray));
 
   /* CudaRtHandler_texture */
   mspHandlers->insert(CUDA_ROUTINE_HANDLER_PAIR(BindTexture));
   mspHandlers->insert(CUDA_ROUTINE_HANDLER_PAIR(BindTexture2D));
-#ifndef CUDART_VERSION
-#error CUDART_VERSION not defined
-#endif
   mspHandlers->insert(CUDA_ROUTINE_HANDLER_PAIR(BindTextureToArray));
   mspHandlers->insert(CUDA_ROUTINE_HANDLER_PAIR(CreateTextureObject));
   mspHandlers->insert(CUDA_ROUTINE_HANDLER_PAIR(GetChannelDesc));
   mspHandlers->insert(CUDA_ROUTINE_HANDLER_PAIR(GetTextureAlignmentOffset));
   mspHandlers->insert(CUDA_ROUTINE_HANDLER_PAIR(GetTextureReference));
   mspHandlers->insert(CUDA_ROUTINE_HANDLER_PAIR(UnbindTexture));
+#endif
 
   /* CudaRtHandler_thread */
   mspHandlers->insert(CUDA_ROUTINE_HANDLER_PAIR(ThreadExit));
   mspHandlers->insert(CUDA_ROUTINE_HANDLER_PAIR(ThreadSynchronize));
 
-  /* CudaRtHandler_version */
 #ifndef CUDART_VERSION
 #error CUDART_VERSION not defined
 #endif
 #if CUDART_VERSION >= 2030
+  /* CudaRtHandler_version */
   mspHandlers->insert(CUDA_ROUTINE_HANDLER_PAIR(DriverGetVersion));
   mspHandlers->insert(CUDA_ROUTINE_HANDLER_PAIR(RuntimeGetVersion));
   mspHandlers->insert(CUDA_ROUTINE_HANDLER_PAIR(FuncSetCacheConfig));
diff --git a/plugins/cudart/backend/CudaRtHandler.h b/plugins/cudart/backend/CudaRtHandler.h
index 1cca403a..d8e4bdb3 100644
--- a/plugins/cudart/backend/CudaRtHandler.h
+++ b/plugins/cudart/backend/CudaRtHandler.h
@@ -60,13 +60,18 @@
 #include "../3rdparty/include/CudaRt_internal.h"
 
 #if (CUDART_VERSION >= 9020)
+
 #if (CUDART_VERSION >= 11000)
 #define __CUDACC__
 #define cudaPushCallConfiguration __cudaPushCallConfiguration
 #endif
+
+#if (CUDART_VERSION < 12000)
 #include "crt/device_functions.h"
 #endif
 
+#endif
+
 //#define DEBUG
 
 /**
@@ -109,6 +114,7 @@ class CudaRtHandler : public gvirtus::backend::Handler {
   const char *GetVar(std::string &handler);
   const char *GetVar(const char *handler);
 
+#if CUDART_VERSION < 12000
   void RegisterTexture(std::string &handler, textureReference *texref);
   void RegisterTexture(const char *handler, textureReference *texref);
   void RegisterSurface(std::string &handler, surfaceReference *surref);
@@ -121,6 +127,7 @@ class CudaRtHandler : public gvirtus::backend::Handler {
   surfaceReference *GetSurface(pointer_t handler);
   surfaceReference *GetSurface(const char *handler);
   const char *GetSurfaceHandler(surfaceReference *texref);
+#endif
 
   const char *GetSymbol(std::shared_ptr<Buffer> in);
 
@@ -168,8 +175,10 @@ class CudaRtHandler : public gvirtus::backend::Handler {
   std::map<std::string, void **> *mpFatBinary;
   std::map<std::string, std::string> *mpDeviceFunction;
   std::map<std::string, std::string> *mpVar;
+#if CUDART_VERSION < 12000
   std::map<std::string, textureReference *> *mpTexture;
   std::map<std::string, surfaceReference *> *mpSurface;
+#endif
   map<std::string, NvInfoFunction>* mapDeviceFunc2InfoFunc;
   map<const void *,std::string>* mapHost2DeviceFunc;
   void *mpShm;
@@ -229,7 +238,7 @@ CUDA_ROUTINE_HANDLER(SetDoubleForDevice);
 CUDA_ROUTINE_HANDLER(SetDoubleForHost);
 CUDA_ROUTINE_HANDLER(SetupArgument);
 
-#if CUDART_VERSION >= 9020
+#if (CUDART_VERSION >= 9020) && (CUDART_VERSION < 12000)
 CUDA_ROUTINE_HANDLER(PushCallConfiguration);
 CUDA_ROUTINE_HANDLER(PopCallConfiguration);
 #endif
@@ -242,8 +251,10 @@ CUDA_ROUTINE_HANDLER(RegisterFunction);
 CUDA_ROUTINE_HANDLER(RegisterVar);
 CUDA_ROUTINE_HANDLER(RegisterSharedVar);
 CUDA_ROUTINE_HANDLER(RegisterShared);
+#if CUDART_VERSION < 12000
 CUDA_ROUTINE_HANDLER(RegisterTexture);
 CUDA_ROUTINE_HANDLER(RegisterSurface);
+#endif
 CUDA_ROUTINE_HANDLER(RegisterSharedMemory);
 CUDA_ROUTINE_HANDLER(RequestSharedMemory);
 
@@ -290,6 +301,7 @@ CUDA_ROUTINE_HANDLER(StreamCreateWithFlags);
 CUDA_ROUTINE_HANDLER(StreamWaitEvent);
 CUDA_ROUTINE_HANDLER(StreamCreateWithPriority);
 
+#if CUDART_VERSION < 12000
 /* CudaRtHandler_texture */
 CUDA_ROUTINE_HANDLER(BindTexture);
 CUDA_ROUTINE_HANDLER(BindTexture2D);
@@ -302,7 +314,7 @@ CUDA_ROUTINE_HANDLER(UnbindTexture);
 
 /* CudaRtHandler_surface */
 CUDA_ROUTINE_HANDLER(BindSurfaceToArray);
-// CUDA_ROUTINE_HANDLER(GetTextureReference);
+#endif
 
 /* CudaRtHandler_thread */
 CUDA_ROUTINE_HANDLER(ThreadExit);
diff --git a/plugins/cudart/backend/CudaRtHandler_device.cpp b/plugins/cudart/backend/CudaRtHandler_device.cpp
index 566c365a..489a8cce 100644
--- a/plugins/cudart/backend/CudaRtHandler_device.cpp
+++ b/plugins/cudart/backend/CudaRtHandler_device.cpp
@@ -358,13 +358,15 @@ CUDA_ROUTINE_HANDLER(IpcOpenEventHandle) {
   Logger logger = Logger::getInstance(LOG4CPLUS_TEXT("IpcOpenEventHandler"));
   CudaRtHandler::setLogLevel(&logger);
 
-  std::shared_ptr<Buffer> out = std::make_shared<Buffer>();
-
   try {
     cudaEvent_t *event = input_buffer->Assign<cudaEvent_t>();
     cudaIpcEventHandle_t handle = input_buffer->Get<cudaIpcEventHandle_t>();
     cudaError_t exit_code = cudaIpcOpenEventHandle(event, handle);
+    std::shared_ptr<Buffer> out = std::make_shared<Buffer>();
+
     out->Add(event);
+    return std::make_shared<Result>(exit_code, out);
+
   } catch (string e) {
     // cerr << e << endl;
     LOG4CPLUS_DEBUG(logger, e);
diff --git a/plugins/cudart/backend/CudaRtHandler_internal.cpp b/plugins/cudart/backend/CudaRtHandler_internal.cpp
index 29d6288a..982e510d 100644
--- a/plugins/cudart/backend/CudaRtHandler_internal.cpp
+++ b/plugins/cudart/backend/CudaRtHandler_internal.cpp
@@ -81,6 +81,7 @@ extern void __cudaRegisterVar(void **fatCubinHandle, char *hostVar,
 extern void __cudaRegisterSharedVar(void **fatCubinHandle, void **devicePtr,
                                     size_t size, size_t alignment, int storage);
 extern void __cudaRegisterShared(void **fatCubinHandle, void **devicePtr);
+#if CUDART_VERSION < 12000
 extern void __cudaRegisterTexture(void **fatCubinHandle,
                                   const textureReference *hostVar,
                                   void **deviceAddress, char *deviceName,
@@ -89,6 +90,7 @@ extern void __cudaRegisterSurface(void **fatCubinHandle,
                                   const surfaceReference *hostVar,
                                   void **deviceAddress, char *deviceName,
                                   int dim, int ext);
+#endif
 }
 
 static bool initialized = false;
@@ -421,6 +423,7 @@ CUDA_ROUTINE_HANDLER(RegisterShared) {
   return std::make_shared<Result>(cudaSuccess);
 }
 
+#if CUDART_VERSION < 12000
 CUDA_ROUTINE_HANDLER(RegisterTexture) {
   try {
     char *handler = input_buffer->AssignString();
@@ -504,15 +507,13 @@ CUDA_ROUTINE_HANDLER(RegisterSurface) {
   return std::make_shared<Result>(cudaSuccess);
 }
 
-#if (CUDART_VERSION >= 9020)
-
+#if (CUDART_VERSION >= 9020) && (CUDART_VERSION < 12000)
 
 #if (CUDART_VERSION >= 11000)
 #define __CUDACC__
 #define cudaPushCallConfiguration __cudaPushCallConfiguration
 #endif
 
-
 #include "crt/device_functions.h"
 #include "CudaRt_internal.h"
 
@@ -573,3 +574,4 @@ CUDA_ROUTINE_HANDLER(PopCallConfiguration) {
 
 }
 #endif
+#endif
diff --git a/plugins/cudart/frontend/CudaRt_internal.cpp b/plugins/cudart/frontend/CudaRt_internal.cpp
index 801e33de..5b662c6a 100644
--- a/plugins/cudart/frontend/CudaRt_internal.cpp
+++ b/plugins/cudart/frontend/CudaRt_internal.cpp
@@ -219,6 +219,7 @@ extern "C" __host__ void __cudaRegisterSharedVar(void **fatCubinHandle,
   CudaRtFrontend::Execute("cudaRegisterSharedVar");
 }
 
+#if CUDART_VERSION < 12000
 extern "C" __host__ void __cudaRegisterTexture(void **fatCubinHandle,
                                                const textureReference *hostVar,
                                                void **deviceAddress,
@@ -255,6 +256,7 @@ extern "C" __host__ void __cudaRegisterSurface(void **fatCubinHandle,
   CudaRtFrontend::AddVariableForArguments(ext);
   CudaRtFrontend::Execute("cudaRegisterSurface");
 }
+#endif
 
 /* */
 
@@ -265,6 +267,7 @@ extern "C" __host__ int __cudaSynchronizeThreads(void **x, void *y) {
   return 0;
 }
 
+#if CUDART_VERSION < 12000
 extern "C" __host__ void __cudaTextureFetch(const void *tex, void *index,
                                             int integer, void *val) {
   // FIXME: implement
@@ -330,3 +333,4 @@ extern "C" cudaError_t CUDARTAPI __cudaPopCallConfiguration( dim3 *gridDim,
     return cudaError;
 }
 #endif
+#endif
diff --git a/plugins/cudnn/backend/CudnnHandler.cpp b/plugins/cudnn/backend/CudnnHandler.cpp
index 4173666d..015d7697 100644
--- a/plugins/cudnn/backend/CudnnHandler.cpp
+++ b/plugins/cudnn/backend/CudnnHandler.cpp
@@ -250,6 +250,7 @@ void CudnnHandler::Initialize(){
     mspHandlers->insert(CUDNN_ROUTINE_HANDLER_PAIR(DropoutBackward));
     mspHandlers->insert(CUDNN_ROUTINE_HANDLER_PAIR(CreateRNNDescriptor));
     mspHandlers->insert(CUDNN_ROUTINE_HANDLER_PAIR(DestroyRNNDescriptor));
+#if CUDNN_VERSION < 9000
 #if CUDNN_VERSION < 8000
     mspHandlers->insert(CUDNN_ROUTINE_HANDLER_PAIR(SetRNNDescriptor_v5));
     //mspHandlers->insert(CUDNN_ROUTINE_HANDLER_PAIR(GetRNNDescriptor_v5));
@@ -285,6 +286,7 @@ void CudnnHandler::Initialize(){
     mspHandlers->insert(CUDNN_ROUTINE_HANDLER_PAIR(RNNBackwardWeights));
     mspHandlers->insert(CUDNN_ROUTINE_HANDLER_PAIR(SetRNNPaddingMode));
     mspHandlers->insert(CUDNN_ROUTINE_HANDLER_PAIR(GetRNNPaddingMode));
+#endif
     mspHandlers->insert(CUDNN_ROUTINE_HANDLER_PAIR(CreateRNNDataDescriptor));
     mspHandlers->insert(CUDNN_ROUTINE_HANDLER_PAIR(DestroyRNNDataDescriptor));
     mspHandlers->insert(CUDNN_ROUTINE_HANDLER_PAIR(SetRNNDataDescriptor));
@@ -4027,7 +4029,10 @@ CUDNN_ROUTINE_HANDLER(DestroyRNNDescriptor){
     return std::make_shared<Result>(cs, out);
 }
 #endif
-#if CUDNN_VERSION >= 6000
+
+#if CUDNN_VERSION >= 9000
+// TODO: implement
+#elif (CUDNN_VERSION >= 6000)
 CUDNN_ROUTINE_HANDLER(SetRNNDescriptor_v6){
     Logger logger = Logger::getInstance(LOG4CPLUS_TEXT("SetRNNDescriptor_v6"));
 
@@ -4198,6 +4203,7 @@ CUDNN_ROUTINE_HANDLER(GetRNNDescriptor_v8) {
 }
 #endif
 
+#if CUDNN_VERSION < 9000
 CUDNN_ROUTINE_HANDLER(SetRNNMatrixMathType){
    Logger logger = Logger::getInstance(LOG4CPLUS_TEXT("SetRNNMatrixMathType"));
    
@@ -4755,6 +4761,7 @@ CUDNN_ROUTINE_HANDLER(GetRNNPaddingMode){
     //cout << " DEBUG - cudnnGetRNNPaddingMode Executed"<<endl;
     return std::make_shared<Result>(cs, out);
 }
+#endif
 
 CUDNN_ROUTINE_HANDLER(CreateRNNDataDescriptor){
      Logger logger = Logger::getInstance(LOG4CPLUS_TEXT("CreateRNNDataDescriptor"));
@@ -4851,6 +4858,7 @@ CUDNN_ROUTINE_HANDLER(GetRNNDataDescriptor){
     return std::make_shared<Result>(cs, out);    
 }
 
+#if CUDNN_VERSION < 9000
 CUDNN_ROUTINE_HANDLER(RNNForwardTrainingEx){
     Logger logger = Logger::getInstance(LOG4CPLUS_TEXT("RNNForwardTrainingEx"));
 
@@ -5271,7 +5279,6 @@ CUDNN_ROUTINE_HANDLER(FindRNNBackwardDataAlgorithmEx){
     return std::make_shared<Result>(cs, out);
 }
 
-
 CUDNN_ROUTINE_HANDLER(GetRNNBackwardWeightsAlgorithmMaxCount){
    Logger logger = Logger::getInstance(LOG4CPLUS_TEXT("GetRNNBackwardWeightsAlgorithmMaxCount"));
 
@@ -5331,6 +5338,7 @@ CUDNN_ROUTINE_HANDLER(FindRNNBackwardWeightsAlgorithmEx){
     //cout << " DEBUG - cudnnFindRNNBackwardWeightsAlgorithmEx Executed"<<endl;
     return std::make_shared<Result>(cs, out);  
 }
+#endif
 
 CUDNN_ROUTINE_HANDLER(CreateSeqDataDescriptor){
     Logger logger = Logger::getInstance(LOG4CPLUS_TEXT("CreateSeqDataDescriptor"));
@@ -5923,6 +5931,7 @@ CUDNN_ROUTINE_HANDLER(GetCTCLossWorkspaceSize){
     return std::make_shared<Result>(cs, out);    
 }
 
+#if CUDNN_VERSION < 9000
 CUDNN_ROUTINE_HANDLER(CreateAlgorithmDescriptor){
     Logger logger = Logger::getInstance(LOG4CPLUS_TEXT("CreateAlgorithmDescriptor"));
 
@@ -6136,6 +6145,7 @@ CUDNN_ROUTINE_HANDLER(RestoreAlgorithm){
     //cout << " DEBUG - cudnnRestoreAlgorithm Executed"<<endl;
     return std::make_shared<Result>(cs);
 }
+#endif
 
 CUDNN_ROUTINE_HANDLER(SetCallback){
      Logger logger = Logger::getInstance(LOG4CPLUS_TEXT("SetCallback"));
diff --git a/plugins/cudnn/backend/CudnnHandler.h b/plugins/cudnn/backend/CudnnHandler.h
index ab43a5c7..7a5b05c2 100644
--- a/plugins/cudnn/backend/CudnnHandler.h
+++ b/plugins/cudnn/backend/CudnnHandler.h
@@ -333,6 +333,7 @@ CUDNN_ROUTINE_HANDLER(DestroyRNNDescriptor);
 CUDNN_ROUTINE_HANDLER(SetRNNDescriptor_v5);
 //CUDNN_ROUTINE_HANDLER(GetRNNDescriptor_v5);
 #endif
+#if CUDNN_VERSION < 9000
 #if CUDNN_VERSION >= 6000
 CUDNN_ROUTINE_HANDLER(SetRNNDescriptor_v6);
 CUDNN_ROUTINE_HANDLER(GetRNNDescriptor_v6);
@@ -363,6 +364,7 @@ CUDNN_ROUTINE_HANDLER(RNNBackwardData);
 CUDNN_ROUTINE_HANDLER(RNNBackwardWeights);
 CUDNN_ROUTINE_HANDLER(SetRNNPaddingMode);
 CUDNN_ROUTINE_HANDLER(GetRNNPaddingMode);
+#endif
 CUDNN_ROUTINE_HANDLER(CreateRNNDataDescriptor);
 CUDNN_ROUTINE_HANDLER(DestroyRNNDataDescriptor);
 CUDNN_ROUTINE_HANDLER(SetRNNDataDescriptor);
diff --git a/plugins/cudnn/frontend/Cudnn.cpp b/plugins/cudnn/frontend/Cudnn.cpp
index c698dc82..4f244f5e 100644
--- a/plugins/cudnn/frontend/Cudnn.cpp
+++ b/plugins/cudnn/frontend/Cudnn.cpp
@@ -3301,6 +3301,8 @@ extern "C" cudnnStatus_t CUDNNWINAPI cudnnDestroyRNNDescriptor(cudnnRNNDescripto
    return CudnnFrontend::GetExitCode();
 }
 
+#if CUDNN_VERSION < 9000
+
 #if CUDNN_VERSION < 8000
 extern "C" cudnnStatus_t CUDNNWINAPI cudnnSetRNNDescriptor_v5(cudnnRNNDescriptor_t rnnDesc,
                                                               int hiddenSize,
@@ -3616,6 +3618,7 @@ extern "C" cudnnStatus_t CUDNNWINAPI cudnnGetRNNProjectionLayers(cudnnHandle_t h
      return CudnnFrontend::GetExitCode();
 }
 
+#if CUDNN_VERSION < 9000
 extern "C" cudnnStatus_t CUDNNWINAPI cudnnCreatePersistentRNNPlan(cudnnRNNDescriptor_t rnnDesc,
                              					  const int minibatch,
                              					  const cudnnDataType_t dataType,
@@ -3657,6 +3660,7 @@ extern "C" cudnnStatus_t CUDNNWINAPI cudnnSetPersistentRNNPlan(cudnnRNNDescripto
 
     return CudnnFrontend::GetExitCode();
 }
+#endif
 
 extern "C" cudnnStatus_t CUDNNWINAPI cudnnGetRNNWorkspaceSize(cudnnHandle_t handle,
                          				     const cudnnRNNDescriptor_t rnnDesc,
@@ -3980,6 +3984,7 @@ extern "C" cudnnStatus_t CUDNNWINAPI cudnnRNNBackwardWeights(cudnnHandle_t handl
       return CudnnFrontend::GetExitCode();
 }
 
+#if CUDNN_VERSION < 9000
 extern "C" cudnnStatus_t CUDNNWINAPI cudnnSetRNNPaddingMode(cudnnRNNDescriptor_t rnnDesc, cudnnRNNPaddingMode_t paddingMode){
 
       CudnnFrontend::Prepare();
@@ -4007,6 +4012,9 @@ extern "C" cudnnStatus_t CUDNNWINAPI cudnnGetRNNPaddingMode(cudnnRNNDescriptor_t
       }
       return CudnnFrontend::GetExitCode();
 }
+#endif
+
+#endif
 
 extern "C" cudnnStatus_t CUDNNWINAPI cudnnCreateRNNDataDescriptor(cudnnRNNDataDescriptor_t *rnnDataDesc){
 
@@ -4333,6 +4341,7 @@ extern "C" cudnnStatus_t CUDNNWINAPI cudnnRNNBackwardWeightsEx(cudnnHandle_t han
       return CudnnFrontend::GetExitCode();
 }
 
+#if CUDNN_VERSION < 9000
 extern "C" cudnnStatus_t CUDNNWINAPI cudnnSetRNNAlgorithmDescriptor(cudnnHandle_t handle, cudnnRNNDescriptor_t rnnDesc, cudnnAlgorithmDescriptor_t algoDesc){
 
      CudnnFrontend::Prepare();
@@ -4346,6 +4355,7 @@ extern "C" cudnnStatus_t CUDNNWINAPI cudnnSetRNNAlgorithmDescriptor(cudnnHandle_
       
       return CudnnFrontend::GetExitCode();
 }
+#endif
 
 extern "C" cudnnStatus_t CUDNNWINAPI cudnnGetRNNForwardInferenceAlgorithmMaxCount(cudnnHandle_t handle, const cudnnRNNDescriptor_t rnnDesc, int *count){
 
@@ -4361,6 +4371,7 @@ extern "C" cudnnStatus_t CUDNNWINAPI cudnnGetRNNForwardInferenceAlgorithmMaxCoun
       return CudnnFrontend::GetExitCode();
 }
 
+#if CUDNN_VERSION < 9000
 extern "C" cudnnStatus_t CUDNNWINAPI cudnnFindRNNForwardInferenceAlgorithmEx(cudnnHandle_t handle,
                                         				     const cudnnRNNDescriptor_t rnnDesc,
                                         				     const int seqLength,
@@ -4418,6 +4429,7 @@ extern "C" cudnnStatus_t CUDNNWINAPI cudnnFindRNNForwardInferenceAlgorithmEx(cud
       }
       return CudnnFrontend::GetExitCode();
 }
+#endif
 
 extern "C" cudnnStatus_t CUDNNWINAPI cudnnGetRNNForwardTrainingAlgorithmMaxCount(cudnnHandle_t handle, const cudnnRNNDescriptor_t rnnDesc, int *count){
 
@@ -4433,6 +4445,7 @@ extern "C" cudnnStatus_t CUDNNWINAPI cudnnGetRNNForwardTrainingAlgorithmMaxCount
       return CudnnFrontend::GetExitCode();
 }
 
+#if CUDNN_VERSION < 9000
 extern "C" cudnnStatus_t CUDNNWINAPI cudnnFindRNNForwardTrainingAlgorithmEx(cudnnHandle_t handle,
                                        				           const cudnnRNNDescriptor_t rnnDesc,
                                        				           const int seqLength,
@@ -4493,6 +4506,7 @@ extern "C" cudnnStatus_t CUDNNWINAPI cudnnFindRNNForwardTrainingAlgorithmEx(cudn
       }
       return CudnnFrontend::GetExitCode();
 }
+#endif
     
 extern "C" cudnnStatus_t CUDNNWINAPI cudnnGetRNNBackwardDataAlgorithmMaxCount(cudnnHandle_t handle, const cudnnRNNDescriptor_t rnnDesc, int *count){
 
@@ -4509,7 +4523,8 @@ extern "C" cudnnStatus_t CUDNNWINAPI cudnnGetRNNBackwardDataAlgorithmMaxCount(cu
       }
       return CudnnFrontend::GetExitCode();
 }
-    
+
+#if CUDNN_VERSION < 9000
 extern "C" cudnnStatus_t CUDNNWINAPI cudnnFindRNNBackwardDataAlgorithmEx(cudnnHandle_t handle,
                                     				        const cudnnRNNDescriptor_t rnnDesc,
                                     					const int seqLength,
@@ -4582,6 +4597,7 @@ extern "C" cudnnStatus_t CUDNNWINAPI cudnnFindRNNBackwardDataAlgorithmEx(cudnnHa
       }
       return CudnnFrontend::GetExitCode();
 }
+#endif
       
 extern "C" cudnnStatus_t CUDNNWINAPI cudnnGetRNNBackwardWeightsAlgorithmMaxCount(cudnnHandle_t handle, const cudnnRNNDescriptor_t rnnDesc, int *count){
      
@@ -4596,6 +4612,7 @@ extern "C" cudnnStatus_t CUDNNWINAPI cudnnGetRNNBackwardWeightsAlgorithmMaxCount
       return CudnnFrontend::GetExitCode();
 }
 
+#if CUDNN_VERSION < 9000
 extern "C" cudnnStatus_t CUDNNWINAPI cudnnFindRNNBackwardWeightsAlgorithmEx(cudnnHandle_t handle,
                                        					    const cudnnRNNDescriptor_t rnnDesc,
                                        					    const int seqLength,
@@ -4647,6 +4664,7 @@ extern "C" cudnnStatus_t CUDNNWINAPI cudnnFindRNNBackwardWeightsAlgorithmEx(cudn
       }
       return CudnnFrontend::GetExitCode();
 }
+#endif
 
 extern "C" cudnnStatus_t CUDNNWINAPI cudnnCreateSeqDataDescriptor(cudnnSeqDataDescriptor_t *seqDataDesc){
 
@@ -5233,6 +5251,7 @@ extern "C" cudnnStatus_t CUDNNWINAPI cudnnGetCTCLossWorkspaceSize(cudnnHandle_t
      return CudnnFrontend::GetExitCode();
 }
 
+#if CUDNN_VERSION < 9000
 extern "C" cudnnStatus_t CUDNNWINAPI cudnnCreateAlgorithmDescriptor(cudnnAlgorithmDescriptor_t *algoDesc){
 
     CudnnFrontend::Prepare();
@@ -5408,6 +5427,7 @@ extern "C" cudnnStatus_t CUDNNWINAPI cudnnRestoreAlgorithm(cudnnHandle_t handle,
     
     return CudnnFrontend::GetExitCode();
 }
+#endif
 
 extern "C" cudnnStatus_t CUDNNWINAPI cudnnSetCallback(unsigned mask, void *udata, cudnnCallback_t fptr){