gpu: sycl: add PReLU primitive implemented using SYCL kernels

oneapi-src · Mar 18, 2023 · 3933b58 · 3933b58
1 parent ed49414
commit 3933b58
Show file tree

Hide file tree

Showing 8 changed files with 856 additions and 4 deletions.
diff --git a/examples/CMakeLists.txt b/examples/CMakeLists.txt
@@ -57,7 +57,6 @@ if(DNNL_SYCL_CUDA)
         ${CMAKE_CURRENT_SOURCE_DIR}/primitives/binary.cpp
         ${CMAKE_CURRENT_SOURCE_DIR}/primitives/lstm.cpp
         ${CMAKE_CURRENT_SOURCE_DIR}/primitives/layer_normalization.cpp
-        ${CMAKE_CURRENT_SOURCE_DIR}/primitives/prelu.cpp
         ${CMAKE_CURRENT_SOURCE_DIR}/primitives/reorder.cpp
         ${CMAKE_CURRENT_SOURCE_DIR}/primitives/shuffle.cpp)
 endif()

diff --git a/src/gpu/nvidia/README.md b/src/gpu/nvidia/README.md
@@ -276,6 +276,11 @@ backward propagation respectively.
 
 * Supported data type are `f32`, `f16`, `bf16` and `s8`.
 
+### PReLU
+The PReLU primitive (Leaky ReLU with a trainable alpha parameter) is implemented
+using SYCL kernels. The primitive supports both forward and backward
+propagations for the data types f32, s32, bf16, f16, s8 and u8.
+
 ### Reorder
 
 The `cudnnTransform` function is the equivalent of oneDNN reorder function.

diff --git a/src/gpu/nvidia/sycl_cuda_engine.cpp b/src/gpu/nvidia/sycl_cuda_engine.cpp
@@ -37,8 +37,8 @@
 #include "gpu/nvidia/sycl_cuda_engine.hpp"
 #include "gpu/nvidia/sycl_cuda_scoped_context.hpp"
 #include "gpu/nvidia/sycl_cuda_stream.hpp"
-
 #include "gpu/sycl/ref_binary.hpp"
+#include "gpu/sycl/ref_prelu.hpp"
 
 namespace dnnl {
 namespace impl {
@@ -215,6 +215,10 @@ constexpr dnnl::impl::impl_list_item_t sycl_cuda_impl_list[] = {
         INSTANCE(cudnn_batch_normalization_fwd_t)
         INSTANCE(cudnn_batch_normalization_bwd_t)
 
+        // PReLU
+        INSTANCE(sycl::ref_prelu_fwd_t)
+        INSTANCE(sycl::ref_prelu_bwd_t)
+
         // Pooling
         INSTANCE(cudnn_pooling_fwd_t)
         INSTANCE(cudnn_pooling_bwd_t)