Skip to content

[ET-VK] Delete unused functions in Context.h, Command.h and Utils.h #4117

New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Closed
wants to merge 1 commit into from
Closed

[ET-VK] Delete unused functions in Context.h, Command.h and Utils.h #4117

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
1 commit
Select commit
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
144 changes: 0 additions & 144 deletions backends/vulkan/runtime/api/Command.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -182,150 +182,6 @@ void CommandBuffer::dispatch(const utils::uvec3& global_workgroup_size) {
state_ = CommandBuffer::State::RECORDING;
}

void CommandBuffer::copy_buffer_to_buffer(
const VulkanBuffer& source,
const VulkanBuffer& destination,
const utils::uvec3& copy_range,
const utils::uvec3& src_offset,
const utils::uvec3& dst_offset) {
VK_CHECK_COND(
state_ == CommandBuffer::State::BARRIERS_INSERTED,
"Vulkan CommandBuffer: called copy_buffer_to_buffer() on a command buffer whose state "
"is not BARRIERS_INSERTED.");

const VkBufferCopy copy_details{
src_offset.data[0u], // srcOffset
dst_offset.data[0u], // dstOffset
copy_range.data[0u], // size
};

vkCmdCopyBuffer(
handle_, source.handle(), destination.handle(), 1u, &copy_details);

state_ = CommandBuffer::State::RECORDING;
}

void CommandBuffer::copy_texture_to_texture(
const VulkanImage& source,
const VulkanImage& destination,
const utils::uvec3& copy_range,
const utils::uvec3& src_offset,
const utils::uvec3& dst_offset) {
VK_CHECK_COND(
state_ == CommandBuffer::State::BARRIERS_INSERTED,
"Vulkan CommandBuffer: called copy_texture_to_texture() on a command buffer whose state "
"is not BARRIERS_INSERTED.");

const VkImageSubresourceLayers src_subresource_layers{
VK_IMAGE_ASPECT_COLOR_BIT, // aspectMask
0u, // mipLevel
0u, // baseArrayLayer
1u, // layerCount
};

const VkImageSubresourceLayers dst_subresource_layers{
VK_IMAGE_ASPECT_COLOR_BIT, // aspectMask
0u, // mipLevel
0u, // baseArrayLayer
1u, // layerCount
};

const VkImageCopy copy_details{
src_subresource_layers, // srcSubresource
create_offset3d(src_offset), // srcOffset
dst_subresource_layers, // dstSubresource
create_offset3d(dst_offset), // dstOffset
create_extent3d(copy_range), // extent
};

vkCmdCopyImage(
handle_,
source.handle(),
source.layout(),
destination.handle(),
destination.layout(),
1u,
&copy_details);

state_ = CommandBuffer::State::RECORDING;
}

void CommandBuffer::copy_texture_to_buffer(
const VulkanImage& source,
const VulkanBuffer& destination,
const utils::uvec3& copy_range,
const utils::uvec3& src_offset,
const utils::uvec3& dst_offset) {
VK_CHECK_COND(
state_ == CommandBuffer::State::BARRIERS_INSERTED,
"Vulkan CommandBuffer: called copy_texture_to_buffer() on a command buffer whose state "
"is not BARRIERS_INSERTED.");

const VkImageSubresourceLayers src_subresource_layers{
VK_IMAGE_ASPECT_COLOR_BIT, // aspectMask
0u, // mipLevel
0u, // baseArrayLayer
1u, // layerCount
};

const VkBufferImageCopy copy_details{
dst_offset.data[0u], // bufferOffset
dst_offset.data[1u], // bufferRowLength
dst_offset.data[2u], // bufferImageHeight
src_subresource_layers, // imageSubresource
create_offset3d(src_offset), // imageOffset
create_extent3d(copy_range), // imageExtent
};

vkCmdCopyImageToBuffer(
handle_,
source.handle(),
source.layout(),
destination.handle(),
1u,
&copy_details);

state_ = CommandBuffer::State::RECORDING;
}

void CommandBuffer::copy_buffer_to_texture(
const VulkanBuffer& source,
const VulkanImage& destination,
const utils::uvec3& copy_range,
const utils::uvec3& src_offset,
const utils::uvec3& dst_offset) {
VK_CHECK_COND(
state_ == CommandBuffer::State::BARRIERS_INSERTED,
"Vulkan CommandBuffer: called copy_buffer_to_texture() on a command buffer whose state "
"is not BARRIERS_INSERTED.");

const VkImageSubresourceLayers dst_subresource_layers{
VK_IMAGE_ASPECT_COLOR_BIT, // aspectMask
0u, // mipLevel
0u, // baseArrayLayer
1u, // layerCount
};

const VkBufferImageCopy copy_details{
src_offset.data[0u], // bufferOffset
src_offset.data[1u], // bufferRowLength
src_offset.data[2u], // bufferImageHeight
dst_subresource_layers, // imageSubresource
create_offset3d(dst_offset), // imageOffset
create_extent3d(copy_range), // imageExtent
};

vkCmdCopyBufferToImage(
handle_,
source.handle(),
destination.handle(),
destination.layout(),
1u,
&copy_details);

state_ = CommandBuffer::State::RECORDING;
}

void CommandBuffer::write_timestamp(VkQueryPool querypool, const uint32_t idx)
const {
VK_CHECK_COND(
Expand Down
31 changes: 1 addition & 30 deletions backends/vulkan/runtime/api/Command.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -39,8 +39,7 @@ class CommandBuffer final {
enum State {
INVALID, // Used to indicate the command buffer is moved from
NEW, // Set during constructor
RECORDING, // Set during call to begin(), dispatch(), and
// copy_*_to_*()
RECORDING, // Set during call to begin() and dispatch()
PIPELINE_BOUND, // Set during call to bind_pipeline()
DESCRIPTORS_BOUND, // Set during call to bind_descriptors()
BARRIERS_INSERTED, // Set during call to insert_barrier()
Expand Down Expand Up @@ -93,34 +92,6 @@ class CommandBuffer final {
void insert_barrier(PipelineBarrier& pipeline_barrier);
void dispatch(const utils::uvec3&);

void copy_buffer_to_buffer(
const VulkanBuffer&,
const VulkanBuffer&,
const utils::uvec3&,
const utils::uvec3&,
const utils::uvec3&);

void copy_texture_to_texture(
const VulkanImage&,
const VulkanImage&,
const utils::uvec3&,
const utils::uvec3&,
const utils::uvec3&);

void copy_texture_to_buffer(
const VulkanImage&,
const VulkanBuffer&,
const utils::uvec3&,
const utils::uvec3&,
const utils::uvec3&);

void copy_buffer_to_texture(
const VulkanBuffer&,
const VulkanImage&,
const utils::uvec3&,
const utils::uvec3&,
const utils::uvec3&);

void write_timestamp(VkQueryPool, const uint32_t) const;
void reset_querypool(VkQueryPool, const uint32_t, const uint32_t) const;

Expand Down
127 changes: 2 additions & 125 deletions backends/vulkan/runtime/api/Context.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -36,8 +36,8 @@ struct ContextConfig final {

//
// Vulkan Context holds onto all relevant Vulkan state as it pertains to our
// use of Vulkan in PyTorch. A Context is associated with one, and only one,
// Adapter as a precursor to multi-GPU support. All Vulkan tensors in PyTorch
// use of Vulkan in PyTorch. A Context is associated with one, and only one,
// Adapter as a precursor to multi-GPU support. All Vulkan tensors in PyTorch
// are associated with a Context to make tensor <-> device affinity explicit.
// The context is currently a global object, but technically it does not need
// to be if we were to make it explicit to the user.
Expand Down Expand Up @@ -200,16 +200,6 @@ class Context final {
const ShaderInfo&,
const utils::uvec3&);

template <class S, class D>
bool submit_copy(
PipelineBarrier&,
const S&,
const D&,
const utils::uvec3&,
const utils::uvec3&,
const utils::uvec3&,
VkFence fence_handle);

template <typename... Arguments>
bool submit_compute_job(
const ShaderInfo&,
Expand Down Expand Up @@ -383,119 +373,6 @@ inline void bind(

} // namespace detail

template <class S, class D>
inline void record_copy(
CommandBuffer& cmd,
const S& source,
const D& destination,
const utils::uvec3& copy_range,
const utils::uvec3& src_offset,
const utils::uvec3& dst_offset) = delete;

template <>
inline void record_copy<VulkanBuffer, VulkanBuffer>(
CommandBuffer& cmd,
const VulkanBuffer& source,
const VulkanBuffer& destination,
const utils::uvec3& copy_range,
const utils::uvec3& src_offset,
const utils::uvec3& dst_offset) {
cmd.copy_buffer_to_buffer(
source, destination, copy_range, src_offset, dst_offset);
}

template <>
inline void record_copy<VulkanImage, VulkanImage>(
CommandBuffer& cmd,
const VulkanImage& source,
const VulkanImage& destination,
const utils::uvec3& copy_range,
const utils::uvec3& src_offset,
const utils::uvec3& dst_offset) {
cmd.copy_texture_to_texture(
source, destination, copy_range, src_offset, dst_offset);
}

template <>
inline void record_copy<VulkanImage, VulkanBuffer>(
CommandBuffer& cmd,
const VulkanImage& source,
const VulkanBuffer& destination,
const utils::uvec3& copy_range,
const utils::uvec3& src_offset,
const utils::uvec3& dst_offset) {
cmd.copy_texture_to_buffer(
source, destination, copy_range, src_offset, dst_offset);
}

template <>
inline void record_copy<VulkanBuffer, VulkanImage>(
CommandBuffer& cmd,
const VulkanBuffer& source,
const VulkanImage& destination,
const utils::uvec3& copy_range,
const utils::uvec3& src_offset,
const utils::uvec3& dst_offset) {
cmd.copy_buffer_to_texture(
source, destination, copy_range, src_offset, dst_offset);
}

/*
Records a GPU data copy into the current command buffer. If the number of
submit_*_job calls exceeds the configured frequency, or if a fence is
provided, then the command buffer is submitted to the GPU for execution.
Returns a bool indicating whether or not the function call resulted in a GPU
queue submission.
*/
template <class S, class D>
inline bool Context::submit_copy(
PipelineBarrier& pipeline_barrier,
const S& source,
const D& destination,
const utils::uvec3& copy_range,
const utils::uvec3& src_offset,
const utils::uvec3& dst_offset,
VkFence fence_handle) {
// If any of the provided arguments does not have memory associated with it,
// then exit early as there is no work to be done. However, if a fence has
// been passed the command buffer is not empty, then the current command
// buffer must still be submitted so that the fence can be signaled.
if (!source || !destination) {
if (fence_handle != VK_NULL_HANDLE && submit_count_ > 0) {
submit_cmd_to_gpu(fence_handle);
return true;
}
return false;
}

// Serialize recording to the shared command buffer. Do not initialize with a
// mutex just yet, since in some cases it will be externally managed.
std::unique_lock<std::mutex> cmd_lock;
// Refer to comments in submit_compute_job for explanation.
if (fence_handle == VK_NULL_HANDLE) {
cmd_lock = std::unique_lock<std::mutex>(cmd_mutex_);
}

set_cmd();

std::string label = "cmd_copy";
report_shader_dispatch_start(label, {0, 0, 0}, {0, 0, 0});

cmd_.insert_barrier(pipeline_barrier);

record_copy(cmd_, source, destination, copy_range, src_offset, dst_offset);

report_shader_dispatch_end();

submit_count_++;
if (fence_handle != VK_NULL_HANDLE ||
submit_count_ >= config_.cmd_submit_frequency) {
submit_cmd_to_gpu(fence_handle);
return true;
}
return false;
}

/*
Records a compute shader dispatch into the current command buffer. If the
number of submit_*_job calls exceeds the configured frequency, or if a fence
Expand Down
13 changes: 0 additions & 13 deletions backends/vulkan/runtime/api/Utils.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -462,19 +462,6 @@ inline int64_t multiply_integers(Iter begin, Iter end) {

} // namespace utils

inline bool operator==(const utils::uvec3& _1, const utils::uvec3& _2) {
return (
_1.data[0u] == _2.data[0u] && _1.data[1u] == _2.data[1u] &&
_1.data[2u] == _2.data[2u]);
}

inline VkOffset3D create_offset3d(const utils::uvec3& offsets) {
return VkOffset3D{
utils::safe_downcast<int32_t>(offsets.data[0u]),
static_cast<int32_t>(offsets.data[1u]),
static_cast<int32_t>(offsets.data[2u])};
}

inline VkExtent3D create_extent3d(const utils::uvec3& extents) {
return VkExtent3D{extents.data[0u], extents.data[1u], extents.data[2u]};
}
Expand Down