[ET-VK] Move files using the Vulkan API to vk_api/

Pull Request resolved: #4125

and move from namespace `api` to `vkapi`. This gave me a major headache in the number of places to update.

This stack orgnaizes ET-VK neatly into three abstraction levels, both by folder and by namespace:
1. `namespace vkcompute` and `graph/`: for operator computation implementation and scheduling.
2. `namespace vkcompute::api` and `graph/api/`: for tensor objects (and other objects wrapping our VulkanBuffer/VulkanImage).
3. `namespace vkcompute::vkapi` and `graph/api/vk_api/`: for direct users of Vulkan API

Additionally, we have
- `namespace vkcompute::utils` and `graph/api/utils/`: for utils used by both namespace `api` and `vkapi`
ghstack-source-id: 232399401
@exported-using-ghexport

Differential Revision: [D59281539](https://our.internmc.facebook.com/intern/diff/D59281539/)

Author: Jorge Pineda

backends/vulkan/runtime/VulkanBackend.cpp

@@ -59,47 +59,47 @@ const uint8_t* get_constant_data_ptr(
   return constant_data + constant_bytes->offset();
 }
 
-api::ScalarType get_scalar_type(const vkgraph::VkDataType& vk_datatype) {
+vkapi::ScalarType get_scalar_type(const vkgraph::VkDataType& vk_datatype) {
   switch (vk_datatype) {
     case vkgraph::VkDataType::BOOL:
-      return api::kBool;
+      return vkapi::kBool;
     case vkgraph::VkDataType::UINT8:
-      return api::kByte;
+      return vkapi::kByte;
     case vkgraph::VkDataType::INT8:
-      return api::kChar;
+      return vkapi::kChar;
     case vkgraph::VkDataType::INT32:
-      return api::kInt;
+      return vkapi::kInt;
     case vkgraph::VkDataType::FLOAT16:
-      return api::kHalf;
+      return vkapi::kHalf;
     case vkgraph::VkDataType::FLOAT32:
-      return api::kFloat;
+      return vkapi::kFloat;
   }
 }
 
-api::StorageType get_storage_type(
+vkapi::StorageType get_storage_type(
     const vkgraph::VkStorageType& vk_storage_type) {
   switch (vk_storage_type) {
     case vkgraph::VkStorageType::BUFFER:
-      return api::kBuffer;
+      return vkapi::kBuffer;
     case vkgraph::VkStorageType::TEXTURE_3D:
-      return api::kTexture3D;
+      return vkapi::kTexture3D;
     case vkgraph::VkStorageType::TEXTURE_2D:
-      return api::kTexture2D;
+      return vkapi::kTexture2D;
     default:
       break;
   }
   VK_THROW("Invalid storage type encountered!");
 }
 
-api::GPUMemoryLayout get_memory_layout(
+vkapi::GPUMemoryLayout get_memory_layout(
     const vkgraph::VkMemoryLayout& vk_memory_layout) {
   switch (vk_memory_layout) {
     case vkgraph::VkMemoryLayout::TENSOR_WIDTH_PACKED:
-      return api::kWidthPacked;
+      return vkapi::kWidthPacked;
     case vkgraph::VkMemoryLayout::TENSOR_HEIGHT_PACKED:
-      return api::kHeightPacked;
+      return vkapi::kHeightPacked;
     case vkgraph::VkMemoryLayout::TENSOR_CHANNELS_PACKED:
-      return api::kChannelsPacked;
+      return vkapi::kChannelsPacked;
     default:
       break;
   }
@@ -115,16 +115,16 @@ GraphConfig get_graph_config(ArrayRef<CompileSpec>& compile_specs) {
     if (strcmp(spec.key, "storage_type_override") == 0) {
       ET_CHECK_MSG(value_size == sizeof(int32_t), "Unexpected value size!");
       int value_as_int = static_cast<int>(getUInt32LE(value_data));
-      api::StorageType storage_type =
-          static_cast<api::StorageType>(value_as_int);
+      vkapi::StorageType storage_type =
+          static_cast<vkapi::StorageType>(value_as_int);
 
       config.set_storage_type_override(storage_type);
     }
     if (strcmp(spec.key, "memory_layout_override") == 0) {
       ET_CHECK_MSG(value_size == sizeof(uint32_t), "Unexpected value size!");
       uint32_t value_as_int = getUInt32LE(value_data);
-      api::GPUMemoryLayout memory_layout =
-          static_cast<api::GPUMemoryLayout>(value_as_int);
+      vkapi::GPUMemoryLayout memory_layout =
+          static_cast<vkapi::GPUMemoryLayout>(value_as_int);
 
       config.set_memory_layout_override(memory_layout);
     }
@@ -171,16 +171,16 @@ class GraphBuilder {
   }
 
   void add_tensor_to_graph(const uint32_t fb_id, VkTensorPtr tensor_fb) {
-    const api::ScalarType& dtype = get_scalar_type(tensor_fb->datatype());
-    api::StorageType storage_type =
+    const vkapi::ScalarType& dtype = get_scalar_type(tensor_fb->datatype());
+    vkapi::StorageType storage_type =
         tensor_fb->storage_type() == vkgraph::VkStorageType::DEFAULT_STORAGE
         ? compute_graph_->suggested_storage_type()
         : get_storage_type(tensor_fb->storage_type());
 
     UIntVector dims_fb = tensor_fb->dims();
     const std::vector<int64_t> dims_vector(dims_fb->cbegin(), dims_fb->cend());
 
-    api::GPUMemoryLayout memory_layout =
+    vkapi::GPUMemoryLayout memory_layout =
         tensor_fb->memory_layout() == vkgraph::VkMemoryLayout::DEFAULT_LAYOUT
         ? compute_graph_->suggested_memory_layout(dims_vector)
         : get_memory_layout(tensor_fb->memory_layout());

backends/vulkan/runtime/api/Context.cpp

@@ -7,7 +7,8 @@
  */
 
 #include <executorch/backends/vulkan/runtime/api/Context.h>
-#include <executorch/backends/vulkan/runtime/api/VkUtils.h>
+
+#include <executorch/backends/vulkan/runtime/api/vk_api/VkUtils.h>
 
 #ifndef VULKAN_DESCRIPTOR_POOL_SIZE
 #define VULKAN_DESCRIPTOR_POOL_SIZE 1024u
@@ -23,7 +24,7 @@ namespace api {
 Context::Context(size_t adapter_i, const ContextConfig& config)
     : config_(config),
       // Important handles
-      adapter_p_(runtime()->get_adapter_p(adapter_i)),
+      adapter_p_(vkapi::runtime()->get_adapter_p(adapter_i)),
       device_(adapter_p_->device_handle()),
       queue_(adapter_p_->request_queue()),
       // Resource pools
@@ -72,8 +73,8 @@ void Context::report_shader_dispatch_start(
         cmd_,
         dispatch_id,
         shader_name,
-        create_extent3d(global_wg_size),
-        create_extent3d(local_wg_size));
+        vkapi::create_extent3d(global_wg_size),
+        vkapi::create_extent3d(local_wg_size));
   }
 }
 
@@ -83,17 +84,17 @@ void Context::report_shader_dispatch_end() {
   }
 }
 
-DescriptorSet Context::get_descriptor_set(
-    const ShaderInfo& shader_descriptor,
+vkapi::DescriptorSet Context::get_descriptor_set(
+    const vkapi::ShaderInfo& shader_descriptor,
     const utils::uvec3& local_workgroup_size,
-    const SpecVarList& additional_constants) {
+    const vkapi::SpecVarList& additional_constants) {
   VkDescriptorSetLayout shader_layout =
       shader_layout_cache().retrieve(shader_descriptor.kernel_layout);
 
   VkPipelineLayout pipeline_layout =
       pipeline_layout_cache().retrieve(shader_layout);
 
-  SpecVarList spec_constants = {
+  vkapi::SpecVarList spec_constants = {
       SV(local_workgroup_size.data[0u]),
       SV(local_workgroup_size.data[1u]),
       SV(local_workgroup_size.data[2u])};
@@ -112,9 +113,9 @@ DescriptorSet Context::get_descriptor_set(
 }
 
 void Context::register_shader_dispatch(
-    const DescriptorSet& descriptors,
-    PipelineBarrier& pipeline_barrier,
-    const ShaderInfo& shader_descriptor,
+    const vkapi::DescriptorSet& descriptors,
+    vkapi::PipelineBarrier& pipeline_barrier,
+    const vkapi::ShaderInfo& shader_descriptor,
     const utils::uvec3& global_workgroup_size) {
   // Adjust the global workgroup size based on the output tile size
   uint32_t global_wg_w = utils::div_up(
@@ -180,12 +181,12 @@ Context* context() {
     try {
       const uint32_t cmd_submit_frequency = 16u;
 
-      const CommandPoolConfig cmd_config{
+      const vkapi::CommandPoolConfig cmd_config{
           32u, // cmdPoolInitialSize
           8u, // cmdPoolBatchSize
       };
 
-      const DescriptorPoolConfig descriptor_pool_config{
+      const vkapi::DescriptorPoolConfig descriptor_pool_config{
           VULKAN_DESCRIPTOR_POOL_SIZE, // descriptorPoolMaxSets
           VULKAN_DESCRIPTOR_POOL_SIZE, // descriptorUniformBufferCount
           VULKAN_DESCRIPTOR_POOL_SIZE, // descriptorStorageBufferCount
@@ -194,7 +195,7 @@ Context* context() {
           32u, // descriptorPileSizes
       };
 
-      const QueryPoolConfig query_pool_config{
+      const vkapi::QueryPoolConfig query_pool_config{
           VULKAN_QUERY_POOL_SIZE, // maxQueryCount
           256u, // initialReserveSize
       };
@@ -206,7 +207,7 @@ Context* context() {
           query_pool_config,
       };
 
-      return new Context(runtime()->default_adapter_i(), config);
+      return new Context(vkapi::runtime()->default_adapter_i(), config);
     } catch (...) {
     }