From 89a278bd1b61c3dc18aaff9ce564e144bc3ece69 Mon Sep 17 00:00:00 2001 From: Rajy Rawther Date: Wed, 1 Nov 2023 18:16:02 -0700 Subject: [PATCH] roc_video_decode class update (#29) * WIP: class implementation * add more definitions * rocvideodecode implementation * formatting fixes * address review comments * rocvideodecode class update * videodec sample app-enable all apis --- samples/videoDecode/videodecode.cpp | 61 ++-- utils/rocvideodecode/roc_video_dec.cpp | 378 ++++++++++++++++--------- utils/rocvideodecode/roc_video_dec.h | 79 +++++- utils/video_demuxer.hpp | 17 ++ 4 files changed, 362 insertions(+), 173 deletions(-) diff --git a/samples/videoDecode/videodecode.cpp b/samples/videoDecode/videodecode.cpp index 1d86951b5e..a84478d143 100644 --- a/samples/videoDecode/videodecode.cpp +++ b/samples/videoDecode/videodecode.cpp @@ -33,13 +33,14 @@ THE SOFTWARE. #include #endif #include "video_demuxer.hpp" -#include "rocdecode.h" +#include "roc_video_dec.h" void ShowHelpAndExit(const char *option = NULL) { std::cout << "Options:" << std::endl << "-i Input File Path - required" << std::endl << "-o Output File Path - dumps output if requested; optional" << std::endl - << "-d GPU device ID (0 for the first device, 1 for the second, etc.); optional; default: 0" << std::endl; + << "-d GPU device ID (0 for the first device, 1 for the second, etc.); optional; default: 0" << std::endl + << "-crop crop rectangle for output (not used when using interopped decoded frame); optional; default: 0" << std::endl; exit(0); } @@ -49,6 +50,9 @@ int main(int argc, char **argv) { int dumpOutputFrames = 0; int isOutputRGB = 0; int deviceId = 0; + Rect crop_rect = {}; + Rect *p_crop_rect = nullptr; + OUTPUT_SURF_MEMORY_TYPE mem_type = OUT_SURFACE_MEM_DEV_INTERNAL; // set to internal // Parse command-line arguments if(argc < 1) { ShowHelpAndExit(); @@ -79,55 +83,64 @@ int main(int argc, char **argv) { deviceId = atoi(argv[i]); continue; } + if (!strcmp(argv[i], "-crop")) { + if (++i == argc || 4 != sscanf(argv[i], "%d,%d,%d,%d", &crop_rect.l, &crop_rect.t, &crop_rect.r, &crop_rect.b)) { + ShowHelpAndExit("-crop"); + } + if ((crop_rect.r - crop_rect.l) % 2 == 1 || (crop_rect.b - crop_rect.t) % 2 == 1) { + std::cout << "output crop rectangle must have width and height of even numbers" << std::endl; + exit(1); + } + p_crop_rect = &crop_rect; + continue; + } ShowHelpAndExit(argv[i]); } - VideoDemuxer demuxer(inputFilePath.c_str()); - //VideoDecode viddec(deviceId); + rocDecVideoCodec rocdec_codec_id = AVCodec2RocDecVideoCodec(demuxer.GetCodecID()); + RocVideoDecoder viddec(deviceId, mem_type, rocdec_codec_id, false, true, p_crop_rect); std::string deviceName, gcnArchName, drmNode; int pciBusID, pciDomainID, pciDeviceID; - /*viddec.getDeviceinfo(deviceName, gcnArchName, pciBusID, pciDomainID, pciDeviceID, drmNode); - std::cout << "info: Using GPU device " << deviceId << ": (drm node: " << drmNode << ") " << deviceName << "[" << gcnArchName << "] on PCI bus " << + viddec.GetDeviceinfo(deviceName, gcnArchName, pciBusID, pciDomainID, pciDeviceID); + std::cout << "info: Using GPU device " << deviceId << deviceName << "[" << gcnArchName << "] on PCI bus " << std::setfill('0') << std::setw(2) << std::right << std::hex << pciBusID << ":" << std::setfill('0') << std::setw(2) << std::right << std::hex << pciDomainID << "." << pciDeviceID << std::dec << std::endl; - std::cout << "info: decoding started, please wait!" << std::endl;*/ + std::cout << "info: decoding started, please wait!" << std::endl; int nVideoBytes = 0, nFrameReturned = 0, nFrame = 0; uint8_t *pVideo = nullptr; uint8_t *pFrame = nullptr; int64_t pts = 0; - //outputImageInfo *pImageInfo; + OutputSurfaceInfo *pSurfInfo; bool bDecodeOutSemiPlanar = false; - uint32_t width, height; - //vcnImageFormat_t subsampling; double totalDecTime = 0; do { auto startTime = std::chrono::high_resolution_clock::now(); demuxer.Demux(&pVideo, &nVideoBytes, &pts); - //nFrameReturned = viddec.decode(pVideo, nVideoBytes, pts); + nFrameReturned = viddec.DecodeFrame(pVideo, nVideoBytes, 0, pts); auto endTime = std::chrono::high_resolution_clock::now(); auto timePerFrame = std::chrono::duration(endTime - startTime).count(); totalDecTime += timePerFrame; - /*if (!nFrame && !viddec.getOutputImageInfo(&pImageInfo)){ - std::cerr << "Error: Failed to get Output Image Info!" << std::endl; + if (!nFrame && !viddec.GetOutputSurfaceInfo(&pSurfInfo)) { + std::cerr << "Error: Failed to get Output Surface Info!" << std::endl; break; - }*/ + } if (dumpOutputFrames) { for (int i = 0; i < nFrameReturned; i++) { - /*pFrame = viddec.getFrame(&pts); - viddec.saveImage(outputFilePath, pFrame, pImageInfo, false); + pFrame = viddec.GetFrame(&pts); + viddec.SaveSurfToFile(outputFilePath, pFrame, pSurfInfo); // release frame - viddec.releaseFrame(pts);*/ + viddec.ReleaseFrame(pts); } } nFrame += nFrameReturned; } while (nVideoBytes); - +#if 0 // is flushing required? // Flush last frames from the decoder if any do { // send null packet to decoder to flush out @@ -135,16 +148,16 @@ int main(int argc, char **argv) { int64_t pts = 0; //nFrameReturned = viddec.decode(pVideo, nVideoBytes, pts); } while (nFrameReturned); - - /*std::cout << "info: Video codec format: " << viddec.getCodecFmtName(viddec.getVcnVideoCodecId()) << std::endl; - std::cout << "info: Video size: [ " << pImageInfo->nOutputWidth << ", " << pImageInfo->nOutputHeight << " ]" << std::endl; - std::cout << "info: Video pix format: " << viddec.getPixFmtName(pImageInfo->chromaFormat) << std::endl; - std::cout << "info: Video Bit depth: " << pImageInfo->nBitDepth << std::endl; +#endif + std::cout << "info: Video codec format: " << viddec.GetCodecFmtName(viddec.GetCodecId()) << std::endl; + std::cout << "info: Video size: [ " << pSurfInfo->output_width << ", " << pSurfInfo->output_height << " ]" << std::endl; + std::cout << "info: Video surface format: " << viddec.GetSurfaceFmtName(pSurfInfo->surface_format) << std::endl; + std::cout << "info: Video Bit depth: " << pSurfInfo->bit_depth << std::endl; std::cout << "info: Total frame decoded: " << nFrame << std::endl; if (!dumpOutputFrames) { std::cout << "info: avg decoding time per frame (ms): " << totalDecTime / nFrame << std::endl; std::cout << "info: avg FPS: " << (nFrame / totalDecTime) * 1000 << std::endl; - }*/ + } return 0; } diff --git a/utils/rocvideodecode/roc_video_dec.cpp b/utils/rocvideodecode/roc_video_dec.cpp index a2e622896d..b04fe3701c 100644 --- a/utils/rocvideodecode/roc_video_dec.cpp +++ b/utils/rocvideodecode/roc_video_dec.cpp @@ -22,9 +22,9 @@ THE SOFTWARE. #include "roc_video_dec.h" -RocVideoDecoder::RocVideoDecoder(int device_id, bool b_use_device_mem, rocDecVideoCodec codec, bool b_low_latency, bool device_frame_pitched, - const Rect *p_crop_rect, const Dim *p_resize_dim, bool extract_user_SEI_Message, int max_width, int max_height, - uint32_t clk_rate, bool force_zero_latency) : device_id_{device_id}, b_use_device_mem_(b_use_device_mem), codec_id_(codec), +RocVideoDecoder::RocVideoDecoder(int device_id, OUTPUT_SURF_MEMORY_TYPE out_mem_type, rocDecVideoCodec codec, bool b_low_latency, bool device_frame_pitched, + const Rect *p_crop_rect, bool extract_user_SEI_Message, int max_width, int max_height, + uint32_t clk_rate, bool force_zero_latency) : device_id_{device_id}, out_mem_type_(out_mem_type), codec_id_(codec), b_low_latency_(b_low_latency), b_device_frame_pitched_(device_frame_pitched), b_extract_sei_message_(extract_user_SEI_Message), max_width_ (max_width), max_height_(max_height), b_force_zero_latency_(force_zero_latency) { @@ -32,7 +32,6 @@ RocVideoDecoder::RocVideoDecoder(int device_id, bool b_use_device_mem, rocDecVid THROW("Failed to initilize the HIP"); } if (p_crop_rect) crop_rect_ = *p_crop_rect; - if (p_resize_dim) resize_dim_ = *p_resize_dim; if (b_extract_sei_message_) { fp_sei_ = fopen("rocdec_sei_message.txt", "wb"); curr_sei_message_ptr_ = new RocdecSeiMessageInfo; @@ -111,6 +110,8 @@ static const char * GetVideoCodecString(rocDecVideoCodec eCodec) { return "Unknown"; } + + /** * @brief function to return the name from codec_id * @@ -122,6 +123,34 @@ const char *RocVideoDecoder::GetCodecFmtName(rocDecVideoCodec codec_id) return GetVideoCodecString(codec_id); } +static const char * GetSurfaceFormatString(rocDecVideoSurfaceFormat surface_format_id) { + static struct { + rocDecVideoSurfaceFormat surf_fmt; + const char *name; + } SurfName [] = { + { rocDecVideoSurfaceFormat_NV12, "NV12" }, + { rocDecVideoSurfaceFormat_P016, "P016" }, + { rocDecVideoSurfaceFormat_YUV444, "YUV444" }, + { rocDecVideoSurfaceFormat_YUV444_16Bit, "YUV444_16Bit" }, + }; + + if (surface_format_id >= rocDecVideoSurfaceFormat_NV12 && surface_format_id <= rocDecVideoSurfaceFormat_YUV444_16Bit) + return SurfName[surface_format_id].name; + else + return "Unknown"; +} + +/** + * @brief function to return the name from surface_format_id + * + * @param surface_format_id - enum for surface format + * @return const char* + */ +const char *RocVideoDecoder::GetSurfaceFmtName(rocDecVideoSurfaceFormat surface_format_id) +{ + return GetSurfaceFormatString(surface_format_id); +} + static const char * GetVideoChromaFormatName(rocDecVideoChromaFormat e_chroma_format) { static struct { rocDecVideoChromaFormat chroma_fmt; @@ -171,6 +200,28 @@ static int GetChromaPlaneCount(rocDecVideoSurfaceFormat surface_format) { return num_planes; } +static void GetSurfaceStrideInternal(rocDecVideoSurfaceFormat surface_format, uint32_t width, uint32_t height, uint32_t *pitch, uint32_t *vstride) { + + switch (surface_format) { + case rocDecVideoSurfaceFormat_NV12: + *pitch = align(width, 256); + *vstride = align(height, 16); + break; + case rocDecVideoSurfaceFormat_P016: + *pitch = align(width, 128) * 2; + *vstride = align(height, 16); + break; + case rocDecVideoSurfaceFormat_YUV444: + *pitch = align(width, 256); + *vstride = align(height, 16); + break; + case rocDecVideoSurfaceFormat_YUV444_16Bit: + *pitch = align(width, 128) * 2; + *vstride = align(height, 16); + break; + } + return; +} /* Return value from HandleVideoSequence() are interpreted as : * 0: fail, 1: succeeded, > 1: override dpb size of parser (set by CUVIDPARSERPARAMS::ulMaxNumDecodeSurfaces while creating parser) @@ -283,46 +334,46 @@ int RocVideoDecoder::HandleVideoSequence(RocdecVideoFormat *pVideoFormat) { videoDecodeCreateInfo.ulMaxWidth = max_width_; videoDecodeCreateInfo.ulMaxHeight = max_height_; - if (!(crop_rect_.r && crop_rect_.b) && !(resize_dim_.w && resize_dim_.h)) { + if (!(crop_rect_.r && crop_rect_.b)) { width_ = pVideoFormat->display_area.right - pVideoFormat->display_area.left; height_ = pVideoFormat->display_area.bottom - pVideoFormat->display_area.top; videoDecodeCreateInfo.ulTargetWidth = pVideoFormat->coded_width; videoDecodeCreateInfo.ulTargetHeight = pVideoFormat->coded_height; } else { - if (resize_dim_.w && resize_dim_.h) { - videoDecodeCreateInfo.display_area.left = pVideoFormat->display_area.left; - videoDecodeCreateInfo.display_area.top = pVideoFormat->display_area.top; - videoDecodeCreateInfo.display_area.right = pVideoFormat->display_area.right; - videoDecodeCreateInfo.display_area.bottom = pVideoFormat->display_area.bottom; - width_ = resize_dim_.w; - height_ = resize_dim_.h; - } - - if (crop_rect_.r && crop_rect_.b) { - videoDecodeCreateInfo.display_area.left = crop_rect_.l; - videoDecodeCreateInfo.display_area.top = crop_rect_.t; - videoDecodeCreateInfo.display_area.right = crop_rect_.r; - videoDecodeCreateInfo.display_area.bottom = crop_rect_.b; - width_ = crop_rect_.r - crop_rect_.l; - height_ = crop_rect_.b - crop_rect_.t; - } + videoDecodeCreateInfo.display_area.left = crop_rect_.l; + videoDecodeCreateInfo.display_area.top = crop_rect_.t; + videoDecodeCreateInfo.display_area.right = crop_rect_.r; + videoDecodeCreateInfo.display_area.bottom = crop_rect_.b; + width_ = crop_rect_.r - crop_rect_.l; + height_ = crop_rect_.b - crop_rect_.t; videoDecodeCreateInfo.ulTargetWidth = width_; videoDecodeCreateInfo.ulTargetHeight = height_; } chroma_height_ = (int)(ceil(height_ * GetChromaHeightFactor(video_surface_format_))); num_chroma_planes_ = GetChromaPlaneCount(video_surface_format_); + if (pVideoFormat->chroma_format == rocDecVideoChromaFormat_Monochrome) num_chroma_planes_ = 0; surface_height_ = videoDecodeCreateInfo.ulTargetHeight; surface_width_ = videoDecodeCreateInfo.ulTargetWidth; - surface_stride_ = align(surface_width_, 256) * byte_per_pixel_; // 256 alignment is enforced for internal VCN surface, keeping the same for ease of memcpy + // 256 alignment is enforced for internal VCN surface, keeping the same for faster memcpy to device memory + surface_stride_ = (out_mem_type_ != OUT_SURFACE_MEM_HOST_COPIED) ? align(surface_width_, 256) * byte_per_pixel_ : surface_width_ * byte_per_pixel_; // fill output_surface_info_ + GetSurfaceStrideInternal(video_surface_format_, surface_width_, surface_height_, &surface_stride_, &surface_vstride_); + chroma_vstride_ = (int)(ceil(surface_vstride_ * GetChromaHeightFactor(video_surface_format_))); output_surface_info_.output_width = surface_width_; output_surface_info_.output_height = surface_height_; - output_surface_info_.output_pitch = surface_stride_; + output_surface_info_.output_pitch = b_device_frame_pitched_? surface_stride_ : surface_width_ * byte_per_pixel_; + output_surface_info_.output_vstride = (out_mem_type_ == OUT_SURFACE_MEM_DEV_INTERNAL) ? surface_vstride_ : surface_height_; output_surface_info_.bit_depth = bitdepth_minus_8_ + 8; output_surface_info_.bytes_per_pixel = byte_per_pixel_; output_surface_info_.surface_format = video_surface_format_; output_surface_info_.num_chroma_planes = num_chroma_planes_; + if (out_mem_type_ == OUT_SURFACE_MEM_DEV_INTERNAL) { + output_surface_info_.output_surface_size_in_bytes = surface_stride_ * (surface_vstride_ + (chroma_vstride_ * num_chroma_planes_)); + } else if (out_mem_type_ == OUT_SURFACE_MEM_DEV_COPIED) + output_surface_info_.output_surface_size_in_bytes = GetFrameSizePitched(); + else + output_surface_info_.output_surface_size_in_bytes = GetFrameSize(); disp_rect_.b = videoDecodeCreateInfo.display_area.bottom; disp_rect_.t = videoDecodeCreateInfo.display_area.top; @@ -428,74 +479,81 @@ int RocVideoDecoder::HandlePictureDisplay(RocdecParserDispInfo *pDispInfo) { if (result == ROCDEC_SUCCESS && (dec_status.decodeStatus == rocDecodeStatus_Error || dec_status.decodeStatus == rocDecodeStatus_Error_Concealed)) { std::cerr << "Decode Error occurred for picture: " << pic_num_in_dec_order_[pDispInfo->picture_index] << std::endl; } - // copy the decoded surface info device or host - uint8_t *p_dec_frame = nullptr; - { + if (out_mem_type_ == OUT_SURFACE_MEM_DEV_INTERNAL) { + DecFrameBuffer dec_frame = { 0 }; + dec_frame.frame_ptr = (uint8_t *)(src_dev_ptr[0]); + dec_frame.pts = pDispInfo->pts; std::lock_guard lock(mtx_vp_frame_); - if ((unsigned)++decoded_frame_cnt_ > vp_frames_.size()) { - // Not enough frames in stock - num_alloced_frames_++; - DecFrameBuffer dec_frame = { 0 }; - if (b_use_device_mem_) { - // allocate based on piched or not - if (b_device_frame_pitched_) - HIP_API_CALL(hipMalloc((void **)&dec_frame.frame_ptr, GetFrameSizePitched())); - else - HIP_API_CALL(hipMalloc((void **)&dec_frame.frame_ptr, GetFrameSize())); + vp_frames_q_.push(dec_frame); + decoded_frame_cnt_++; + } else { + // copy the decoded surface info device or host + uint8_t *p_dec_frame = nullptr; + { + std::lock_guard lock(mtx_vp_frame_); + // if not enough frames in stock, allocate + if ((unsigned)++decoded_frame_cnt_ > vp_frames_.size()) { + num_alloced_frames_++; + DecFrameBuffer dec_frame = { 0 }; + if (out_mem_type_ == OUT_SURFACE_MEM_DEV_COPIED) { + // allocate based on piched or not + if (b_device_frame_pitched_) + HIP_API_CALL(hipMalloc((void **)&dec_frame.frame_ptr, GetFrameSizePitched())); + else + HIP_API_CALL(hipMalloc((void **)&dec_frame.frame_ptr, GetFrameSize())); + } else { + dec_frame.frame_ptr = new uint8_t[GetFrameSize()]; + } + dec_frame.pts = pDispInfo->pts; + vp_frames_.push_back(dec_frame); } - else{ - dec_frame.frame_ptr = new uint8_t[GetFrameSize()]; - } - dec_frame.pts = pDispInfo->pts; - vp_frames_.push_back(dec_frame); + p_dec_frame = vp_frames_[decoded_frame_cnt_ - 1].frame_ptr; } - p_dec_frame = vp_frames_[decoded_frame_cnt_ - 1].frame_ptr; - } + // Copy luma data + int dst_pitch = b_device_frame_pitched_? surface_stride_ : width_*byte_per_pixel_; + if (out_mem_type_ == OUT_SURFACE_MEM_DEV_COPIED) { + if (src_pitch[0] == dst_pitch) { + int luma_size = src_pitch[0] * height_; + HIP_API_CALL(hipMemcpyDtoDAsync(p_dec_frame, src_dev_ptr[0], luma_size, hip_stream_)); + } else { + // use 2d copy to copy an ROI + HIP_API_CALL(hipMemcpy2DAsync(p_dec_frame, dst_pitch, src_dev_ptr[0], src_pitch[0], width_ * byte_per_pixel_, height_, hipMemcpyDeviceToDevice, hip_stream_)); + } + } else + HIP_API_CALL(hipMemcpy2DAsync(p_dec_frame, width_ * byte_per_pixel_, src_dev_ptr[0], src_pitch[0], width_ * byte_per_pixel_, height_, hipMemcpyDeviceToHost, hip_stream_)); - // Copy luma data - int dst_pitch = b_device_frame_pitched_? surface_stride_ : width_*byte_per_pixel_; - if (b_use_device_mem_) { - if (src_pitch[0] == dst_pitch) { - int luma_size = src_pitch[0] * height_; - HIP_API_CALL(hipMemcpyDtoDAsync(p_dec_frame, src_dev_ptr[0], luma_size, hip_stream_)); - }else { - // use 2d copy to copy an ROI - HIP_API_CALL(hipMemcpy2DAsync(p_dec_frame, dst_pitch, src_dev_ptr[0], src_pitch[0], width_*byte_per_pixel_, height_, hipMemcpyDeviceToDevice, hip_stream_)); - } - } - else - HIP_API_CALL(hipMemcpy2DAsync(p_dec_frame, width_*byte_per_pixel_, src_dev_ptr[0], src_pitch[0], width_*byte_per_pixel_, height_, hipMemcpyDeviceToHost, hip_stream_)); - - // Copy chroma plane ( ) - // rocDec output gives pointer to luma and chroma pointers seperated for the decoded frame - uint8_t *p_frame_uv = p_dec_frame + dst_pitch * height_; - if (b_use_device_mem_) { - if (src_pitch[1] == dst_pitch) { - int chroma_size = chroma_height_ * dst_pitch; - HIP_API_CALL(hipMemcpyDtoDAsync(p_frame_uv, src_dev_ptr[1], chroma_size, hip_stream_)); - }else { - // use 2d copy to copy an ROI - HIP_API_CALL(hipMemcpy2DAsync(p_frame_uv, dst_pitch, src_dev_ptr[1], src_pitch[1], width_*byte_per_pixel_, chroma_height_, hipMemcpyDeviceToDevice, hip_stream_)); - } - }else - HIP_API_CALL(hipMemcpy2DAsync(p_frame_uv, dst_pitch, src_dev_ptr[1], src_pitch[1], width_*byte_per_pixel_, chroma_height_, hipMemcpyDeviceToHost, hip_stream_)); - - if (num_chroma_planes_ == 2) { - uint8_t *p_frame_uv = p_dec_frame + dst_pitch * height_*2; - if (b_use_device_mem_) { - if (src_pitch[2] == dst_pitch) { + // Copy chroma plane ( ) + // rocDec output gives pointer to luma and chroma pointers seperated for the decoded frame + uint8_t *p_frame_uv = p_dec_frame + dst_pitch * height_; + if (out_mem_type_ == OUT_SURFACE_MEM_DEV_COPIED) { + if (src_pitch[1] == dst_pitch) { int chroma_size = chroma_height_ * dst_pitch; - HIP_API_CALL(hipMemcpyDtoDAsync(p_frame_uv, src_dev_ptr[2], chroma_size, hip_stream_)); + HIP_API_CALL(hipMemcpyDtoDAsync(p_frame_uv, src_dev_ptr[1], chroma_size, hip_stream_)); }else { // use 2d copy to copy an ROI - HIP_API_CALL(hipMemcpy2DAsync(p_frame_uv, dst_pitch, src_dev_ptr[2], src_pitch[2], width_*byte_per_pixel_, chroma_height_, hipMemcpyDeviceToDevice, hip_stream_)); + HIP_API_CALL(hipMemcpy2DAsync(p_frame_uv, dst_pitch, src_dev_ptr[1], src_pitch[1], width_ * byte_per_pixel_, chroma_height_, hipMemcpyDeviceToDevice, hip_stream_)); } - }else - HIP_API_CALL(hipMemcpy2DAsync(p_frame_uv, dst_pitch, src_dev_ptr[2], src_pitch[2], width_*byte_per_pixel_, chroma_height_, hipMemcpyDeviceToHost, hip_stream_)); + } else + HIP_API_CALL(hipMemcpy2DAsync(p_frame_uv, dst_pitch, src_dev_ptr[1], src_pitch[1], width_ * byte_per_pixel_, chroma_height_, hipMemcpyDeviceToHost, hip_stream_)); + + if (num_chroma_planes_ == 2) { + uint8_t *p_frame_uv = p_dec_frame + dst_pitch * (height_ + chroma_height_); + if (out_mem_type_ == OUT_SURFACE_MEM_DEV_COPIED) { + if (src_pitch[2] == dst_pitch) { + int chroma_size = chroma_height_ * dst_pitch; + HIP_API_CALL(hipMemcpyDtoDAsync(p_frame_uv, src_dev_ptr[2], chroma_size, hip_stream_)); + } else { + // use 2d copy to copy an ROI + HIP_API_CALL(hipMemcpy2DAsync(p_frame_uv, dst_pitch, src_dev_ptr[2], src_pitch[2], width_*byte_per_pixel_, chroma_height_, hipMemcpyDeviceToDevice, hip_stream_)); + } + }else + HIP_API_CALL(hipMemcpy2DAsync(p_frame_uv, dst_pitch, src_dev_ptr[2], src_pitch[2], width_*byte_per_pixel_, chroma_height_, hipMemcpyDeviceToHost, hip_stream_)); + } + + HIP_API_CALL(hipStreamSynchronize(hip_stream_)); + ROCDEC_API_CALL(rocDecUnMapVideoFrame(roc_decoder_, src_dev_ptr[0])); } - HIP_API_CALL(hipStreamSynchronize(hip_stream_)); - ROCDEC_API_CALL(rocDecUnMapVideoFrame(roc_decoder_, src_dev_ptr[0])); return 1; } @@ -534,99 +592,143 @@ int RocVideoDecoder::GetSEIMessage(RocdecSeiMessageInfo *pSEIMessageInfo) { int RocVideoDecoder::DecodeFrame(const uint8_t *data, size_t size, int pkt_flags, int64_t pts) { - int decoded_frame_cnt_ = 0, decoded_frame_cnt_ret_ = 0; - RocdecSourceDataPacket packet = { 0 }; - packet.payload = data; - packet.payload_size = size; - packet.flags = pkt_flags | ROCDEC_PKT_TIMESTAMP; - packet.pts = pts; - if (!data || size == 0) { - packet.flags |= ROCDEC_PKT_ENDOFSTREAM; - } - ROCDEC_API_CALL(rocDecParseVideoData(rocdec_parser_, &packet)); + if (data && size) { + int decoded_frame_cnt_ = 0, decoded_frame_cnt_ret_ = 0; + RocdecSourceDataPacket packet = { 0 }; + packet.payload = data; + packet.payload_size = size; + packet.flags = pkt_flags | ROCDEC_PKT_TIMESTAMP; + packet.pts = pts; + if (!data || size == 0) { + packet.flags |= ROCDEC_PKT_ENDOFSTREAM; + } + ROCDEC_API_CALL(rocDecParseVideoData(rocdec_parser_, &packet)); - return decoded_frame_cnt_; + return decoded_frame_cnt_; + } else + return 0; } uint8_t* RocVideoDecoder::GetFrame(int64_t *pts) { if (decoded_frame_cnt_ > 0) { std::lock_guard lock(mtx_vp_frame_); decoded_frame_cnt_--; - if (pts) *pts = vp_frames_[decoded_frame_cnt_ret_].pts; - return vp_frames_[decoded_frame_cnt_ret_++].frame_ptr; + if (out_mem_type_ == OUT_SURFACE_MEM_DEV_INTERNAL && !vp_frames_q_.empty()) { + DecFrameBuffer *fb = &vp_frames_q_.front(); + if (pts) *pts = fb->pts; + return fb->frame_ptr; + }else { + if (pts) *pts = vp_frames_[decoded_frame_cnt_ret_].pts; + return vp_frames_[decoded_frame_cnt_ret_++].frame_ptr; + } } return nullptr; } -#if 0 // may be needed for future +/** + * @brief function to release frame after use by the application: Only used with "OUT_SURFACE_MEM_DEV_INTERNAL" + * + * @param pTimestamp - timestamp of the frame to be released (unmapped) + * @return true - success + * @return false - falied + */ -void RocVideoDecoder::SaveImage(std::string output_file_name, void *dev_mem, OutputImageInfo *image_info, bool is_output_RGB) { +bool RocVideoDecoder::ReleaseFrame(int64_t pTimestamp) { + if (out_mem_type_ != OUT_SURFACE_MEM_DEV_INTERNAL) + return true; // nothing to do + // only needed when using internal mapped buffer + if (!vp_frames_q_.empty()) { + std::lock_guard lock(mtx_vp_frame_); + DecFrameBuffer *fb = &vp_frames_q_.front(); + void *mapped_frame_ptr = fb->frame_ptr; + + if (pTimestamp != fb->pts) { + std::cerr << "Decoded Frame is released out of order" << std::endl; + return false; + } + ROCDEC_API_CALL(rocDecUnMapVideoFrame(roc_decoder_, mapped_frame_ptr)); + // pop decoded frame + vp_frames_q_.pop(); + } + return true; +} + + +void RocVideoDecoder::SaveSurfToFile(std::string output_file_name, void *surf_mem, OutputSurfaceInfo *surf_info) { uint8_t *hst_ptr = nullptr; - uint64_t output_image_size = image_info->output_image_size_in_bytes; - if (hst_ptr == nullptr) { - hst_ptr = new uint8_t [output_image_size]; - } - hipError_t hip_status = hipSuccess; - hip_status = hipMemcpyDtoH((void *)hst_ptr, dev_mem, output_image_size); - if (hip_status != hipSuccess) { - std::cout << "ERROR: hipMemcpyDtoH failed! (" << hip_status << ")" << std::endl; - delete [] hst_ptr; - return; - } + uint64_t output_image_size = surf_info->output_surface_size_in_bytes; + if (surf_info->mem_type == OUT_SURFACE_MEM_DEV_INTERNAL || surf_info->mem_type == OUT_SURFACE_MEM_DEV_COPIED) { + if (hst_ptr == nullptr) { + hst_ptr = new uint8_t [output_image_size]; + } + hipError_t hip_status = hipSuccess; + hip_status = hipMemcpyDtoH((void *)hst_ptr, surf_mem, output_image_size); + if (hip_status != hipSuccess) { + std::cout << "ERROR: hipMemcpyDtoH failed! (" << hip_status << ")" << std::endl; + delete [] hst_ptr; + return; + } + } else + hst_ptr = (uint8_t *)surf_mem; + - // no RGB dump if the surface type is YUV400 - if (image_info->chroma_format == ROCDEC_FMT_YUV400 && is_output_RGB) { - return; - } uint8_t *tmp_hst_ptr = hst_ptr; if (fp_out_ == nullptr) { fp_out_ = fopen(output_file_name.c_str(), "wb"); } if (fp_out_) { - int img_width = image_info->output_width; - int img_height = image_info->output_height; - int output_image_stride = image_info->output_h_stride; - if (img_width * image_info->bytes_per_pixel == output_image_stride && img_height == image_info->output_v_stride) { + int img_width = surf_info->output_width; + int img_height = surf_info->output_height; + int output_stride = surf_info->output_pitch; + if (img_width * surf_info->bytes_per_pixel == output_stride && img_height == surf_info->output_vstride) { fwrite(hst_ptr, 1, output_image_size, fp_out_); } else { - uint32_t width = is_output_RGB ? image_info->output_width * 3 : image_info->output_width; - if (image_info->bit_depth == 8) { - for (int i = 0; i < image_info->output_height; i++) { + uint32_t width = surf_info->output_width; + if (surf_info->bit_depth == 8) { + for (int i = 0; i < surf_info->output_height; i++) { fwrite(tmp_hst_ptr, 1, width, fp_out_); - tmp_hst_ptr += output_image_stride; + tmp_hst_ptr += output_stride; } - if (!is_output_RGB) { - // dump chroma - uint8_t *uv_hst_ptr = hst_ptr + output_image_stride * image_info->output_v_stride; - for (int i = 0; i < img_height >> 1; i++) { + // dump chroma + uint8_t *uv_hst_ptr = hst_ptr + output_stride * surf_info->output_vstride; + for (int i = 0; i < chroma_height_; i++) { + fwrite(uv_hst_ptr, 1, width, fp_out_); + uv_hst_ptr += output_stride; + } + if (num_chroma_planes_ == 2) { + uint8_t *v_hst_ptr = hst_ptr + output_stride * (surf_info->output_vstride + chroma_vstride_); + for (int i = 0; i < chroma_height_; i++) { fwrite(uv_hst_ptr, 1, width, fp_out_); - uv_hst_ptr += output_image_stride; + v_hst_ptr += output_stride; } } - } else if (image_info->bit_depth > 8 && image_info->bit_depth <= 16 ) { + + } else if (surf_info->bit_depth > 8 && surf_info->bit_depth <= 16 ) { for (int i = 0; i < img_height; i++) { - fwrite(tmp_hst_ptr, 1, width * image_info->bytes_per_pixel, fp_out_); - tmp_hst_ptr += output_image_stride; + fwrite(tmp_hst_ptr, 1, width * surf_info->bytes_per_pixel, fp_out_); + tmp_hst_ptr += output_stride; } - if (!is_output_RGB) { - // dump chroma - uint8_t *uv_hst_ptr = hst_ptr + output_image_stride * image_info->output_v_stride; - for (int i = 0; i < img_height >> 1; i++) { - fwrite(uv_hst_ptr, 1, width * image_info->bytes_per_pixel, fp_out_); - uv_hst_ptr += output_image_stride; + // dump chroma + uint8_t *uv_hst_ptr = hst_ptr + output_stride * surf_info->output_vstride; + for (int i = 0; i < chroma_height_; i++) { + fwrite(uv_hst_ptr, 1, width * surf_info->bytes_per_pixel, fp_out_); + uv_hst_ptr += output_stride; + } + if (num_chroma_planes_ == 2) { + uint8_t *v_hst_ptr = hst_ptr + output_stride * (surf_info->output_vstride + chroma_vstride_); + for (int i = 0; i < chroma_height_; i++) { + fwrite(uv_hst_ptr, 1, width, fp_out_); + v_hst_ptr += output_stride; } } } } } - if (hst_ptr != nullptr) { + if (hst_ptr && (surf_info->mem_type != OUT_SURFACE_MEM_HOST_COPIED)) { delete [] hst_ptr; - hst_ptr = nullptr; - tmp_hst_ptr = nullptr; } } -#endif void RocVideoDecoder::GetDeviceinfo(std::string &device_name, std::string &gcn_arch_name, int &pci_bus_id, int &pci_domain_id, int &pci_device_id) { device_name = hip_dev_prop_.name; diff --git a/utils/rocvideodecode/roc_video_dec.h b/utils/rocvideodecode/roc_video_dec.h index ca61f8dfa0..adfa9f5d6f 100644 --- a/utils/rocvideodecode/roc_video_dec.h +++ b/utils/rocvideodecode/roc_video_dec.h @@ -30,10 +30,12 @@ THE SOFTWARE. #include #include #include +#include +#include +#include #include #include "rocdecode.h" #include "rocparser.h" -#include "commons.h" #define MAX_FRAME_NUM 16 @@ -42,6 +44,35 @@ typedef enum{ SEI_TYPE_USER_DATA_UNREGISTERED = 5 }SEI_H264_HEVC_PAYLOAD_TYPE; +typedef enum { + OUT_SURFACE_MEM_DEV_INTERNAL = 0, /**< Internal interopped decoded surface memory(original mapped decoded surface) */ + OUT_SURFACE_MEM_DEV_COPIED = 1, /**< decoded output will be copied to a separate device memory (the user doesn't need to call release) **/ + OUT_SURFACE_MEM_HOST_COPIED = 2 /**< decoded output will be copied to a separate host memory (the user doesn't need to call release) **/ +} OUTPUT_SURF_MEMORY_TYPE; + +#define TOSTR(X) std::to_string(static_cast(X)) +#define STR(X) std::string(X) + +#if DBGINFO +#define INFO(X) std::clog << "[INF] " << " {" << __func__ <<"} " << " " << X << std::endl; +#else +#define INFO(X) ; +#endif +#define ERR(X) std::cerr << "[ERR] " << " {" << __func__ <<"} " << " " << X << std::endl; + + +class rocVideoDecodeException : public std::exception { +public: + + explicit rocVideoDecodeException(const std::string& message):_message(message){} + virtual const char* what() const throw() override { + return _message.c_str(); + } +private: + std::string _message; +}; + +#define THROW(X) throw rocVideoDecodeException(" { "+std::string(__func__)+" } " + X); #define ROCDEC_API_CALL( rocDecAPI ) \ do { \ @@ -71,10 +102,6 @@ struct Rect { int l, t, r, b; }; -struct Dim { - int w, h; -}; - static inline int align(int value, int alignment) { return (value + alignment - 1) & ~(alignment - 1); } @@ -89,11 +116,13 @@ typedef struct OutputSurfaceInfoType { uint32_t output_width; /**< Output width of decoded surface*/ uint32_t output_height; /**< Output height of decoded surface*/ uint32_t output_pitch; /**< Output pitch in bytes of luma plane, chroma pitch can be inferred based on chromaFormat*/ + uint32_t output_vstride; /**< Output vertical stride in case of using internal mem pointer **/ uint32_t bytes_per_pixel; /**< Output BytesPerPixel of decoded image*/ uint32_t bit_depth; /**< Output BitDepth of the image*/ uint32_t num_chroma_planes; /**< Output Chroma number of planes*/ uint64_t output_surface_size_in_bytes; /**< Output Image Size in Bytes; including both luma and chroma planes*/ rocDecVideoSurfaceFormat surface_format; /**< Chroma format of the decoded image*/ + OUTPUT_SURF_MEMORY_TYPE mem_type; /**< Output mem_type of the surface*/ } OutputSurfaceInfo; class RocVideoDecoder { @@ -108,18 +137,18 @@ class RocVideoDecoder { * @param b_low_latency * @param device_frame_pitched * @param p_crop_rect - * @param p_resize_dim * @param extract_user_SEI_Message * @param max_width * @param max_height * @param clk_rate * @param force_zero_latency */ - RocVideoDecoder(int device_id, bool b_use_device_mem, rocDecVideoCodec codec, bool b_low_latency, bool device_frame_pitched, - const Rect *p_crop_rect, const Dim *p_resize_dim, bool extract_user_SEI_Message, int max_width, int max_height, - uint32_t clk_rate, bool force_zero_latency); + RocVideoDecoder(int device_id, OUTPUT_SURF_MEMORY_TYPE out_mem_type, rocDecVideoCodec codec, bool b_low_latency, bool device_frame_pitched, + const Rect *p_crop_rect, bool extract_user_SEI_Message = false, int max_width = 0, int max_height = 0, + uint32_t clk_rate = 1000, bool force_zero_latency = false); ~RocVideoDecoder(); + rocDecVideoCodec GetCodecId() { return codec_id_; } /** * @brief Get the output frame width */ @@ -177,6 +206,15 @@ class RocVideoDecoder { * @return std::string */ const char *GetCodecFmtName(rocDecVideoCodec codec_id); + + /** + * @brief function to return the name from surface_format_id + * + * @param surface_format_id - enum for surface format + * @return const char* + */ + const char *GetSurfaceFmtName(rocDecVideoSurfaceFormat surface_format_id); + /** * @brief Get the pointer to the Output Image Info * @@ -201,6 +239,15 @@ class RocVideoDecoder { */ uint8_t* GetFrame(int64_t *pts); + /** + * @brief function to release frame after use by the application: Only used with "OUT_SURFACE_MEM_DEV_INTERNAL" + * + * @param pTimestamp - timestamp of the frame to be released (unmapped) + * @return true - success + * @return false - falied + */ + bool ReleaseFrame(int64_t pTimestamp); + /** * @brief utility function to save image to a file * @@ -221,6 +268,15 @@ class RocVideoDecoder { * @param pci_device_id */ void GetDeviceinfo(std::string &device_name, std::string &gcn_arch_name, int &pci_bus_id, int &pci_domain_id, int &pci_device_id); + + /** + * @brief Helper function to dump decoded output surface to file + * + * @param output_file_name - Output file name + * @param dev_mem - pointer to surface memory + * @param surf_info - surface info + */ + void SaveSurfToFile(std::string output_file_name, void *surf_mem, OutputSurfaceInfo *surf_info); private: int decoder_session_id_; // Decoder session identifier. Used to gather session level stats. @@ -280,7 +336,7 @@ class RocVideoDecoder { int device_id_; RocdecVideoParser rocdec_parser_ = nullptr; rocDecDecoderHandle roc_decoder_ = nullptr; - bool b_use_device_mem_ = true; + OUTPUT_SURF_MEMORY_TYPE out_mem_type_ = OUT_SURFACE_MEM_DEV_INTERNAL; bool b_extract_sei_message_ = false; bool b_low_latency_ = true; bool b_force_zero_latency_ = true; @@ -308,13 +364,14 @@ class RocVideoDecoder { uint32_t num_chroma_planes_; uint32_t num_components_; uint32_t surface_stride_; + uint32_t surface_vstride_, chroma_vstride_; // vertical stride between planes: used when using internal dev memory size_t surface_size_; OutputSurfaceInfo output_surface_info_; std::mutex mtx_vp_frame_; std::vector vp_frames_; // vector of decoded frames + std::queue vp_frames_q_; Rect disp_rect_ = {}; Rect crop_rect_ = {}; - Dim resize_dim_ = {}; FILE *fp_sei_ = NULL; FILE *fp_out_ = NULL; }; \ No newline at end of file diff --git a/utils/video_demuxer.hpp b/utils/video_demuxer.hpp index 5c6f97fcce..99d8473030 100644 --- a/utils/video_demuxer.hpp +++ b/utils/video_demuxer.hpp @@ -33,6 +33,7 @@ extern "C" { #include } +#include "rocdecode.h" /*! * \file * \brief The AMD Video Demuxer for rocDecode Library. @@ -50,6 +51,7 @@ class VideoDemuxer { virtual ~StreamProvider() {} virtual int GetData(uint8_t *buf, int buf_size) = 0; }; + AVCodecID GetCodecID() { return av_video_codec_id_; }; VideoDemuxer(const char *input_file_path) : VideoDemuxer(CreateFmtContextUtil(input_file_path)) {} VideoDemuxer(StreamProvider *stream_provider) : VideoDemuxer(CreateFmtContextUtil(stream_provider)) {av_io_ctx_ = av_fmt_input_ctx_->pb;} ~VideoDemuxer(); @@ -270,3 +272,18 @@ AVFormatContext *VideoDemuxer::CreateFmtContextUtil(const char *input_file_path) int VideoDemuxer::ReadPacket(void *data, uint8_t *buf, int buf_size) { return ((StreamProvider *)data)->GetData(buf, buf_size); } + +static inline rocDecVideoCodec AVCodec2RocDecVideoCodec(AVCodecID av_codec) { + switch (av_codec) { + case AV_CODEC_ID_MPEG1VIDEO : return rocDecVideoCodec_MPEG1; + case AV_CODEC_ID_MPEG2VIDEO : return rocDecVideoCodec_MPEG2; + case AV_CODEC_ID_MPEG4 : return rocDecVideoCodec_MPEG4; + case AV_CODEC_ID_H264 : return rocDecVideoCodec_H264; + case AV_CODEC_ID_HEVC : return rocDecVideoCodec_HEVC; + case AV_CODEC_ID_VP8 : return rocDecVideoCodec_VP8; + case AV_CODEC_ID_VP9 : return rocDecVideoCodec_VP9; + case AV_CODEC_ID_MJPEG : return rocDecVideoCodec_JPEG; + case AV_CODEC_ID_AV1 : return rocDecVideoCodec_AV1; + default : return rocDecVideoCodec_NumCodecs; + } +} \ No newline at end of file