Skip to content
Closed
Show file tree
Hide file tree
Changes from 1 commit
Commits
File filter

Filter by extension

Filter by extension

Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
Prev Previous commit
Next Next commit
refactor
  • Loading branch information
AlexanderKalistratov committed Jun 27, 2023
commit 809047daf6040b00fa492ba1441e280f18da9af7
Original file line number Diff line number Diff line change
Expand Up @@ -33,20 +33,16 @@ def initialize(
default_rng.random((batch, in_channels, in_height, in_width)).astype(
dtype
),
# np.ones((batch, in_channels, in_height, in_width)).astype(dtype),
np.zeros((batch, out_channels, out_height, out_width)).astype(dtype),
# np.zeros((kernel_height, kernel_width, 2, out_height, out_width)).astype(dtype),
2
* default_rng.random(
(kernel_height, kernel_width, 2, out_height, out_width)
).astype(dtype)
- 1,
# default_rng.random((out_channels, in_channels, kernel_height, kernel_width)).astype(dtype),
np.ones(
(out_channels, in_channels, kernel_height, kernel_width)
).astype(dtype),
default_rng.random(out_channels).astype(dtype),
# np.ones((out_channels,)).astype(dtype),
np.zeros(
(in_channels, kernel_height, kernel_width, out_height, out_width)
).astype(dtype),
Expand Down

This file was deleted.

This file was deleted.

Original file line number Diff line number Diff line change
@@ -1,19 +1,16 @@
cmake_minimum_required(VERSION 3.23)

set(py_module_name _deformable_convolution_sycl)
pybind11_add_module(${py_module_name}
MODULE
deformable_convolution_sycl/impl.cpp
)
find_package(MKL CONFIG REQUIRED)

# target_include_directories(${py_module_name} PUBLIC ${Dpctl_INCLUDE_DIRS} $)
target_compile_options(${py_module_name} PUBLIC $<TARGET_PROPERTY:MKL::MKL,INTERFACE_COMPILE_OPTIONS>)
target_include_directories(${py_module_name} PUBLIC ${Dpctl_INCLUDE_DIRS} $<TARGET_PROPERTY:MKL::MKL,INTERFACE_INCLUDE_DIRECTORIES>)
target_link_libraries(${py_module_name} PUBLIC $<LINK_ONLY:MKL::MKL> ${MKL_SYCL})

message(STATUS "${MKL_IMPORTED_TARGETS}")
message(STATUS "${INTERFACE_COMPILE_OPTIONS}")
message(STATUS "${INTERFACE_INCLUDE_DIRECTORIES}")

file(RELATIVE_PATH py_module_dest ${CMAKE_SOURCE_DIR} ${CMAKE_CURRENT_SOURCE_DIR})
install(TARGETS ${py_module_name}
DESTINATION ${py_module_dest}/deformable_convolution_sycl
Expand Down
Original file line number Diff line number Diff line change
Expand Up @@ -119,15 +119,11 @@ inline auto deform_input(cl::sycl::queue &queue,
out_width, kh, kw, 0, h, w) +
w * stride_x + (kw - k_w_m) * dilation_x -
(pad_x - k_w_m);
// auto offset_y = h*stride_y + (kh - k_h_m)*dilation_y - (pad_y -
// k_h_m); auto offset_x = w*stride_x + (kw - k_w_m)*dilation_x - (pad_x
// - k_w_m);

auto _input =
get_ptr_3d(input, in_channels, in_height, in_width, c, 0, 0);

_output[w] = bilinear(_input, in_height, in_width, offset_y, offset_x);
// _output[w] = offset_y;
});
}

Expand Down Expand Up @@ -197,11 +193,11 @@ void deformable_convolution_b1_impl(cl::sycl::queue &queue,
auto efill = output_fill_with_bias(queue, output, out_shape, bias);
auto egemm = gemm(queue, transpose::N, transpose::N, /*transpose a, b*/
out_c, out_h * out_w, in_c * ker_h * ker_w, /*m, n, k*/
1, /*alpha*/
weights, in_c * ker_h * ker_w, /*a, lda*/
tmp, out_h * out_w, /*b, ldb*/
1, /*beta*/
output, out_h * out_w, /*c, ldc*/
1, /*alpha*/
weights, in_c * ker_h * ker_w, /*a, lda*/
tmp, out_h * out_w, /*b, ldb*/
1, /*beta*/
output, out_h * out_w, /*c, ldc*/
{edeform, efill} /*events*/);
egemm.wait();
}
Expand Down
Loading