AI/rocm-systems
2
0
Çatalla 0
Kod Konular Değişiklik İstekleri İşlemler Paketler Projeler Sürüm Viki Aktivite

Modify the memory pool access for batch decode submission (#50)

Kaynağa Gözat
Bu işleme şunda yer alıyor:
Aryan Salmanpour
2024-08-14 16:45:05 -04:00
işlemeyi yapan: GitHub
ebeveyn b68b9ba8ea
işleme 7129a01f88
3 değiştirilmiş dosya ile 93 ekleme ve 30 silme
Yama Dosyasını İndir Diff Dosyasını İndir Tüm dosyaları genişlet Tüm dosyaları daralt
src/rocjpeg_decoder.cpp
+2
Dosyayı Görüntüle
@@ -172,6 +172,7 @@ RocJpegStatus RocJpegDecoder::Decode(RocJpegStreamHandle jpeg_stream_handle, con
break;
}
CHECK_ROCJPEG(jpeg_vaapi_decoder_.SetSurfaceAsIdle(current_surface_id));
CHECK_HIP(hipStreamSynchronize(hip_stream_));
return ROCJPEG_STATUS_SUCCESS;
}
@@ -268,6 +269,7 @@ RocJpegStatus RocJpegDecoder::DecodeBatched(RocJpegStreamHandle *jpeg_streams, i
default:
break;
}
CHECK_ROCJPEG(jpeg_vaapi_decoder_.SetSurfaceAsIdle(current_surface_id));
}
}
src/rocjpeg_vaapi_decoder.cpp
+61 -21
Dosyayı Görüntüle
@@ -110,25 +110,32 @@ RocJpegStatus RocJpegVaapiMemoryPool::AddPoolEntry(uint32_t surface_format, cons
if (entries.size() < max_pool_size_) {
entries.push_back(pool_entry);
} else {
if (entries.front().va_context_id != 0) {
CHECK_VAAPI(vaDestroyContext(va_display_, entries.front().va_context_id));
entries.front().va_context_id = 0;
}
if (!entries.front().va_surface_ids.empty()) {
CHECK_VAAPI(vaDestroySurfaces(va_display_, entries.front().va_surface_ids.data(), entries.front().va_surface_ids.size()));
std::fill(entries.front().va_surface_ids.begin(), entries.front().va_surface_ids.end(), 0);
}
if (!entries.front().hip_interops.empty()) {
for(auto& hip_interop_entry : entries.front().hip_interops) {
if (hip_interop_entry.hip_mapped_device_mem != nullptr)
CHECK_HIP(hipFree(hip_interop_entry.hip_mapped_device_mem));
if (hip_interop_entry.hip_ext_mem != nullptr)
CHECK_HIP(hipDestroyExternalMemory(hip_interop_entry.hip_ext_mem));
memset((void*)&hip_interop_entry, 0, sizeof(hip_interop_entry));
auto it = std::find_if(entries.begin(), entries.end(), [](const RocJpegVaapiMemPoolEntry& entry) {return entry.entry_status == kIdle;});
if (it != entries.end()) {
auto index = std::distance(entries.begin(), it);
if (entries[index].va_context_id != 0) {
CHECK_VAAPI(vaDestroyContext(va_display_, entries[index].va_context_id));
entries[index].va_context_id = 0;
}
if (!entries[index].va_surface_ids.empty()) {
CHECK_VAAPI(vaDestroySurfaces(va_display_, entries[index].va_surface_ids.data(), entries[index].va_surface_ids.size()));
std::fill(entries[index].va_surface_ids.begin(), entries[index].va_surface_ids.end(), 0);
}
if (!entries[index].hip_interops.empty()) {
for(auto& hip_interop_entry : entries[index].hip_interops) {
if (hip_interop_entry.hip_mapped_device_mem != nullptr)
CHECK_HIP(hipFree(hip_interop_entry.hip_mapped_device_mem));
if (hip_interop_entry.hip_ext_mem != nullptr)
CHECK_HIP(hipDestroyExternalMemory(hip_interop_entry.hip_ext_mem));
memset((void*)&hip_interop_entry, 0, sizeof(hip_interop_entry));
}
}
entries.erase(it);
entries.push_back(pool_entry);
} else {
ERR("cannot find an idle entry in the the memory pool!");
return ROCJPEG_STATUS_INVALID_PARAMETER;
}
entries.erase(entries.begin());
entries.push_back(pool_entry);
}
return ROCJPEG_STATUS_SUCCESS;
}
@@ -143,12 +150,13 @@ RocJpegStatus RocJpegVaapiMemoryPool::AddPoolEntry(uint32_t surface_format, cons
* @return The matching `RocJpegVaapiMemPoolEntry` if found, or a default-initialized entry if not found.
*/
RocJpegVaapiMemPoolEntry RocJpegVaapiMemoryPool::GetEntry(uint32_t surface_format, uint32_t image_width, uint32_t image_height, uint32_t num_surfaces) {
for (const auto& entry : mem_pool_[surface_format]) {
if (entry.image_width == image_width && entry.image_height == image_height && entry.va_surface_ids.size() == num_surfaces) {
for (auto& entry : mem_pool_[surface_format]) {
if (entry.image_width == image_width && entry.image_height == image_height && entry.va_surface_ids.size() == num_surfaces && entry.entry_status == kIdle) {
entry.entry_status = kBusy;
return entry;
}
}
return {0, 0, 0 , {}, {}};
return {0, 0, 0 , kIdle, {}, {}};
}
bool RocJpegVaapiMemoryPool::FindSurfaceId(VASurfaceID surface_id) {
@@ -229,6 +237,19 @@ RocJpegStatus RocJpegVaapiMemoryPool::GetHipInteropMem(VASurfaceID surface_id, H
ERR("the surface_id: " + TOSTR(surface_id) + " was not found in the memory pool!");
return ROCJPEG_STATUS_INVALID_PARAMETER;
}
bool RocJpegVaapiMemoryPool::SetSurfaceAsIdle(VASurfaceID surface_id) {
for (auto& pair : mem_pool_) {
for (auto& entry : pair.second) {
if (std::find(entry.va_surface_ids.begin(), entry.va_surface_ids.end(), surface_id) != entry.va_surface_ids.end()) {
entry.entry_status = kIdle;
return true;
}
}
}
return false;
}
/**
* @brief Constructs a RocJpegVappiDecoder object.
*
@@ -356,7 +377,7 @@ RocJpegStatus RocJpegVappiDecoder::InitializeDecoder(std::string device_name, st
INFO("WARNING: didn't find the vcn jpeg spec for " + gcn_arch_name_base_temp + " using the default setting");
current_vcn_jpeg_spec_.num_jpeg_cores = 1;
}
vaapi_mem_pool_->SetPoolSize(current_vcn_jpeg_spec_.num_jpeg_cores * 2);
vaapi_mem_pool_->SetPoolSize(current_vcn_jpeg_spec_.num_jpeg_cores + 1);
return ROCJPEG_STATUS_SUCCESS;
}
@@ -562,6 +583,7 @@ RocJpegStatus RocJpegVappiDecoder::SubmitDecode(const JpegStreamParameters *jpeg
mem_pool_entry.va_context_id = va_context_id;
mem_pool_entry.hip_interops.resize(1);
surface_id = mem_pool_entry.va_surface_ids[0];
mem_pool_entry.entry_status = kBusy;
CHECK_ROCJPEG(vaapi_mem_pool_->AddPoolEntry(surface_pixel_format, mem_pool_entry));
} else {
surface_id = mem_pool_entry.va_surface_ids[0];
@@ -683,6 +705,7 @@ RocJpegStatus RocJpegVappiDecoder::SubmitDecodeBatched(JpegStreamParameters *jpe
}
mem_pool_entry.va_context_id = va_context_id;
mem_pool_entry.hip_interops.resize(indices.size());
mem_pool_entry.entry_status = kBusy;
CHECK_ROCJPEG(vaapi_mem_pool_->AddPoolEntry(key.pixel_format, mem_pool_entry));
} else {
for (int i = 0; i < mem_pool_entry.va_surface_ids.size(); i++) {
@@ -888,4 +911,21 @@ void RocJpegVappiDecoder::GetDrmNodeOffset(std::string device_name, uint8_t devi
break;
}
}
}
/**
* @brief Sets a VASurfaceID as idle.
*
* This function sets the specified VASurfaceID as idle in the RocJpegVappiDecoder's vaapi_mem_pool.
* If the surface cannot be set as idle, it returns ROCJPEG_STATUS_INVALID_PARAMETER.
*
* @param surface_id The VASurfaceID to set as idle.
* @return RocJpegStatus The status of the operation. Returns ROCJPEG_STATUS_SUCCESS if successful,
* or ROCJPEG_STATUS_INVALID_PARAMETER if the surface cannot be set as idle.
*/
RocJpegStatus RocJpegVappiDecoder::SetSurfaceAsIdle(VASurfaceID surface_id) {
if (!vaapi_mem_pool_->SetSurfaceAsIdle(surface_id)) {
return ROCJPEG_STATUS_INVALID_PARAMETER;
}
return ROCJPEG_STATUS_SUCCESS;
}
src/rocjpeg_vaapi_decoder.h
+30 -9
Dosyayı Görüntüle
@@ -95,23 +95,26 @@ struct HipInteropDeviceMem {
};
/**
* @brief Structure representing an entry in the RocJpegVappiMemPool.
*
* This structure holds information about an image in the memory pool used by the RocJpegVappiDecoder.
* It includes the image width and height, the VASurfaceID and VAContextID associated with the image,
* and the HipInteropDeviceMem for interoperation with HIP.
* @brief Defines the enumeration MemPoolEntryStatus.
*/
typedef enum {
kIdle = 0,
kBusy = 1,
} MemPoolEntryStatus;
/**
* @brief Structure representing an entry in the RocJpegVappiMemPool.
* @struct RocJpegVaapiMemPoolEntry
* @brief Structure representing an entry in the RocJpegVaapiMemPool.
*
* This structure holds information about an image in the memory pool used by the RocJpegVappiDecoder.
* It includes the image width and height, the VASurfaceID and VAContextID associated with the image,
* and the HipInteropDeviceMem for interoperation with HIP.
* This structure holds information about a memory pool entry used by the RocJpegVaapiDecoder.
* It contains the image width and height, the VA context ID, the entry status, an array of VA surface IDs,
* and an array of HipInteropDeviceMem objects.
*/
struct RocJpegVaapiMemPoolEntry {
uint32_t image_width;
uint32_t image_height;
VAContextID va_context_id;
MemPoolEntryStatus entry_status;
std::vector<VASurfaceID> va_surface_ids;
std::vector<HipInteropDeviceMem> hip_interops;
};
@@ -181,6 +184,16 @@ class RocJpegVaapiMemoryPool {
*/
RocJpegStatus GetHipInteropMem(VASurfaceID surface_id, HipInteropDeviceMem& hip_interop);
/**
* @brief Sets a VASurfaceID as idle.
*
* This function sets the specified VASurfaceID as idle, indicating that it is available for reuse.
*
* @param surface_id The VASurfaceID to set as idle.
* @return true if the VASurfaceID was successfully set as idle, false otherwise.
*/
bool SetSurfaceAsIdle(VASurfaceID surface_id);
private:
VADisplay va_display_; // The VADisplay associated with the memory pool.
uint32_t max_pool_size_; // The maximum pool size of the memory pool (mem_pool_) per entry.
@@ -302,6 +315,14 @@ public:
* @return The current VCN JPEG specification.
*/
const VcnJpegSpec& GetCurrentVcnJpegSpec() const {return current_vcn_jpeg_spec_;}
/**
* Sets the specified VASurfaceID as idle.
*
* @param surface_id The VASurfaceID to set as idle.
* @return The status of the operation.
*/
RocJpegStatus SetSurfaceAsIdle(VASurfaceID surface_id);
private:
int device_id_; // The ID of the device
int drm_fd_; // The file descriptor for the DRM device