Code modified for CUDA Graph in the sampleINTAPI is throwing error

trivedi.nagaraj · January 7, 2024, 8:59am

Hi all, I tried modifying the sample code file sampleINT8API.cpp from the directory /usr/src/tensorrt/samples/sampleINT8API/

It is throwing the following error

[01/07/2024-03:22:44] [I] [TRT] [MemUsageSnapshot] ExecutionContext creation end: CPU 1373 MiB, GPU 7177 MiB
[01/07/2024-03:22:44] [I] Started capturing CUDA graph

[01/07/2024-03:22:44] [E] [TRT] 1: [blobInfo.cpp::getHostScale::803] Error Code 1: Cuda Runtime (operation not permitted when stream is capturing)
[01/07/2024-03:22:44] [F] [TRT] [defaultAllocator.cpp::free::85] Error Code 1: Cuda Runtime (operation not permitted when stream is capturing)
[01/07/2024-03:22:44] [I] [TRT] [MemUsageChange] Init cuBLAS/cuBLASLt: CPU +0, GPU +0, now: CPU 1372, GPU 7177 (MiB)
[01/07/2024-03:22:44] [F] [TRT] [defaultAllocator.cpp::free::85] Error Code 1: Cuda Runtime (operation not permitted when stream is capturing)

I have modified only one API sample::Logger::TestResult SampleINT8API::infer()

For your reference I have attached the file containing contents of this entire function. My modifications are under the macro
ifdef TRT_DEBUG and endif to differentiate my modifications from the original code. The TRT_DEBUG macro is defined in the beginning of the source file.

Thanks and Regards

Nagaraj Trivedi

code_modification_cuda_graph_capture.txt (2.3 KB)

SivaRamaKrishnaNV · January 7, 2024, 9:35am

Dear @trivedi.nagaraj,
From past topics, I am assuming you are looking for Performing inference Resnet50 model with sample Input and check layer fusion info and impact of cudaGraph.
May I know using useCudaGraph flag in trtexec did not help your need? This flag enables cudaGraph APIs calls. It can be noticed using nsys trace.
I know you have some issue with trtexec with loadInput at The trt exec could not predict the image properly with resNet50.onnx model . But I believe it could be issue with Jetpack version or input data or input parameters to trtexec. Please double check that.
trtexec is the best way to have all your needs in a single sample with limited modifications.

trivedi.nagaraj · January 7, 2024, 10:23am

Hi SivaRamaKirshna, yes I am looking for ResNet50 model inference with layer fusion, INT8 precision and CUDA graph.
But trtexec inference is not working as it is not predicting correctly.

Also installation of JetPack 5.0 version is giving errors for me. Is it possible that we can have a google meet and I will share you my screen and you can help me in installing the JetPack 5.0

Thanks and Regards

Nagaraj Trivedi

SivaRamaKrishnaNV · January 7, 2024, 10:35am

The command I provide works for FP32. Is it the actually issue? Are you using additional int8 flags with trtexec

trivedi.nagaraj · January 7, 2024, 11:14am

Hi SivaramKrishna, not matter whether I use INT8 or FP32 it is not predicting correctly.

Thanks and Regards

Nagaraj Trivedi

trivedi.nagaraj · January 8, 2024, 6:40am

Hi ShivaramaKrishna, if you fix the cuda graph error which I am getting in the sampleINTAPI then I am done with what I require. Because through this I have already achieved INT8, Layer fusion. Now the only item which is pending is CUDA graph. If you fix this problem then I may require trtexec now.
But before that get it clarified with can CUDA graph be used when there is an asynchronous copying (copy between host and device memory) and context->enqueuev2() is used.

Thanks and Regards

Nagaraj Trivedi

trivedi.nagaraj · January 8, 2024, 11:06am

Hi ShivaRamaKrishna, I have fixed this issue and is working fine without throwing any errors.
Here is the output from my logs which I launched cuda graph 2 times

[01/08/2024-05:43:26] [I] Started capturing CUDA graph

[01/08/2024-05:43:26] [I] Ending capturing CUDA graph

[01/08/2024-05:43:26] [I] Launching CUDA graph

[01/08/2024-05:43:26] [I] SampleINT8API result: Detected:
[01/08/2024-05:43:26] [I] [1] airliner
[01/08/2024-05:43:26] [I] [2] warplane
[01/08/2024-05:43:26] [I] [3] projectile
[01/08/2024-05:43:26] [I] [4] space shuttle
[01/08/2024-05:43:26] [I] [5] wing
[01/08/2024-05:43:26] [I] SampleINT8API result: Detected:
[01/08/2024-05:43:26] [I] [1] airliner
[01/08/2024-05:43:26] [I] [2] warplane
[01/08/2024-05:43:26] [I] [3] projectile
[01/08/2024-05:43:26] [I] [4] space shuttle
[01/08/2024-05:43:26] [I] [5] wing

The fix is, we need to call the context->enqueuev2() once before beginning the cuda graph capture.

I got this solution from the nvidia github link when I started searching how to implement the cuda graph in c++. From this link below I found the solution

github.com/NVIDIA/TensorRT

tensor rt cuda graph use with plugin.

opened 06:14AM - 29 Sep 20 UTC

closed 09:42AM - 02 Nov 20 UTC

kycocotree

Plugins triaged

## Description ## Environment **TensorRT Version**: 7.1.3.4 **GPU …Type**: GTX 1080 **Nvidia Driver Version**: 440.100 **CUDA Version**: 10.2 **CUDNN Version**: 8.0 **Operating System + Version**: Ubuntu 18.04 **Python Version (if applicable)**: **TensorFlow Version (if applicable)**: **PyTorch Version (if applicable)**: **Baremetal or Container (which commit + image + tag)**: ## Relevant Files ## Steps To Reproduce  Hi, I made a tensorrt engine by adding a plugin to the yolov3 model. I want to launch a graph using cudaStreamCapture function. (bellow code, (BuildGraph).) However, when capturing a graph, 900 cudaError occurs in the cuda kernel function added as a plugin. How can I create a kernel function made for plugin use to capture it in graph? bool TRTContext::BuildGraph(trt_device::TrtCudaStream& stream) { /* https://docs.nvidia.com/deeplearning/tensorrt/api/c_api/classnvinfer1_1_1_i_execution_context.html#ac7a5737264c2b7860baef0096d961f5a Note: Calling enqueueV2() with a stream in CUDA graph capture mode has a known issue. If dynamic shapes are used, the first enqueueV2() call after a setInputShapeBinding() call will cause failure in stream capture due to resource allocation. Please call enqueueV2() once before capturing the graph. */ bool captured = true; if (m_bExplicitBatch_) { m_pExecutionContext_->enqueueV2(&m_vecDeviceBindings_[0], stream.get(), nullptr); } captured = m_cudaGraph_.beginCapture(stream); if (captured) { if (m_bExplicitBatch_) { if (!m_pExecutionContext_->enqueueV2(&m_vecDeviceBindings_[0], stream.get(), nullptr)) { // CHECK(cuerr); captured = false; } printf("[%s] Use enqueueV2\n", __FUNCTION__); } else { if (!m_pExecutionContext_->enqueue(m_iMaxBatchSize_, &m_vecDeviceBindings_[0], stream.get(), nullptr)) { // CHECK(cuerr); captured = false; } } if (captured) { captured = m_cudaGraph_.endCapture(stream); } } return captured; } In .cu file int cuda::decode (int batch_size, const void *const *inputs, void *outputs, size_t height, size_t width, size_t num_count, size_t num_classes, const std::vector<float> &anchors, size_t stride, size_t grid_w, size_t grid_h, void *workspace, size_t workspace_size, cudaStream_t stream) { int anchor_size = anchors.size() / num_count; int anchors_byte_size = sizeof(float) * anchors.size(); void* anchors_d = nullptr; CUDA_CHECK(cudaMalloc(&anchors_d, anchors_byte_size)); CUDA_CHECK(cudaMemcpyAsync(anchors_d, anchors.data(), anchors_byte_size, cudaMemcpyHostToDevice, stream)); cudaStreamSynchronize(stream); int input_elements_size = grid_w * grid_h * num_count * (1 + LOCATIONS + num_classes); int output_elements_size = (grid_w * grid_h * num_count * sizeof(Detection) / sizeof(float)); CUDA_CHECK(cudaMemset(static_cast<float*>(outputs), 0, output_elements_size * batch_size * sizeof(float))); int num_threads = 1024; int num_elements = grid_h * grid_w; if (num_elements < num_threads) num_threads = num_elements; int num_blocks = 1; for (int b = 0; b < batch_size; b++) { const float* input = static_cast<const float *>(inputs[0]) + (b * input_elements_size); float* output = static_cast<float *>(outputs) + (b * output_elements_size); const int num_per_thread = num_elements / num_threads; if (num_per_thread >= 1) num_blocks = num_per_thread; //printf("num_threads: %d, num_per_thread: %d, num_elements:%d\n", num_threads, num_per_thread, num_elements); decode_kernel <<< num_blocks, num_threads, 0, stream >>> (num_per_thread, input, output, static_cast<float*>(anchors_d), anchor_size, grid_w, grid_h, width, height, num_classes, num_count, num_elements); } cudaStreamSynchronize(stream); CUDA_CHECK(cudaFree(anchors_d)); return cudaSuccess; }

In the above link I found this information
/*
TensorRT: nvinfer1::IExecutionContext Class Reference
Note:
Calling enqueueV2() with a stream in CUDA graph capture mode has a known issue.
If dynamic shapes are used, the first enqueueV2() call after a setInputShapeBinding() call will cause failure in stream capture due to resource allocation.
Please call enqueueV2() once before capturing the graph.
*/

Thank you for all the support and help you have provided me till now.

I will further continue with activities

Modify the code to perform N number of inferences, where N can be provided as one of the command line inputs to it
I want to see this difference in execution speed with and without cuda graph.
Use the DLA core option to find the energy efficiency

I may further require your help in modifying this code to

Make it generic to convert and infer any models other than ResNet50.onnx
Make use of DLA core to find the energy efficiency.

Please help me in that.

Thanks and Regards

Nagaraj Trivedi

trivedi.nagaraj · January 8, 2024, 12:14pm

Hi ShivaRamaKrishna, while executing this binary (sampleINT8API), I have found one warning message
[W] [TRT] Calibrator is not being used. Users must provide dynamic range for all tensors that are not Int32.

Does it mean that model is not converted with INT8 precision ? or It has converted the model with INT8 precision but weights are still considered as 32 bits, please clarify.

If your answer is “It has not converted the model to INT8 precision” then let me know where to find the calibrator and if it is not present then how to create it.

Thanks and Regards

Nagaraj Trivedi

SivaRamaKrishnaNV · January 8, 2024, 2:29pm

Dear @trivedi.nagaraj,
The INT8 model is generated using calibrator or by using dynamic range info provided using external file. You can check this in build() function.
If we use calibrator, it finds the active Tensor range for a set of input set of images used in calibration process.

trivedi.nagaraj · January 8, 2024, 3:29pm

Hi ShivaRamaKrishnan, thank you for the response. I am more interested in inferencing

Once with calibrator file
Once with providing the dynamic range. Let me know how it can be done.
I have read the build() method and it is mentioned here
// Enable INT8 model. Required to set custom per-tensor dynamic range or INT8 Calibration
config->setFlag(BuilderFlag::kINT8);
// Mark calibrator as null. As user provides dynamic range for each tensor, no calibrator is required
config->setInt8Calibrator(nullptr);
// force layer to execute with required precision
setLayerPrecision(network);

// set INT8 Per Tensor Dynamic range
if (!setDynamicRange(network))
{
sample::gLogError << “Unable to set per-tensor dynamic range.” << std::endl;
return sample::Logger::TestResult::kFAILED;
}

And when I inferred it with --verbose option
and found this information
[01/08/2024-10:13:13] [I] Layer: node_of_gpu_0/res2_0_branch2b_bn_1. Precision: INT8
[01/08/2024-10:13:13] [I] Tensor: gpu_0/res2_0_branch2b_bn_1. OutputType: INT8

[01/08/2024-10:13:13] [I] If dynamic range for a tensor is missing, TensorRT will run inference assuming dynamic range for the tensor as optional.
[01/08/2024-10:13:13] [I] If dynamic range for a tensor is required then inference will fail. Follow README.md to generate missing per-tensor dynamic range.
[01/08/2024-10:13:13] [V] [TRT] Setting dynamic range for gpu_0/data_0 to [-1.00024,1.00024]

[01/08/2024-10:13:13] [I] Per Tensor Dynamic Range Values for the Network:
[01/08/2024-10:13:13] [I] Tensor: (Unnamed Layer* 180) [Softmax]_output. Max Absolute Dynamic Range: 0.0303731
[01/08/2024-10:13:13] [I] Tensor: (Unnamed Layer* 179) [Shuffle]_output. Max Absolute Dynamic Range: 6.46343
[01/08/2024-10:13:13] [I] Tensor: (Unnamed Layer* 177) [Shuffle]_output. Max Absolute Dynamic Range: 0.0365279
[01/08/2024-10:13:13] [I] Tensor: (Unnamed Layer* 176) [Constant]_output. Max Absolute Dynamic Range: 0.0365279
[01/08/2024-10:13:13] [I] Tensor: (Unnamed Layer* 175) [Fully Connected]_output. Max Absolute Dynamic Range: 6.40009
[01/08/2024-10:13:13] [I] Tensor: (Unnamed Layer* 174) [Constant]_output. Max Absolute Dynamic Range: 0.443716

Does it mean that though both calibration file and dynamic range are not mentioned but still it is converted to INT8. Please clarify it.

The next clarification I have is how to generate the calibrator for making use in config->setInt8Calibrator(nullptr); , instead of nullptr

Thanks and Regards

Nagaraj Trivedi

trivedi.nagaraj · January 9, 2024, 7:40am

Hi ShiavaRamaKrishna, in continuation with my previous query/clarifications I few more listed below.

How to modify this code to infer a batch of input images (let us say a batch size of 32, 64, 128, 256). This information is required for me to verify how the batch size is impacting model inference speed.
As of now it is reading the input image file of type .ppm, how to modify it to work with .jpg, .jpeg or .dat files

Please clarify these doubts.

Thanks and Regards

Nagaraj Trivedi

trivedi.nagaraj · January 9, 2024, 1:43pm

Hi ShivaRamaKrishna, please update me for the queries I have raised. I tried converting the .jpg image to the .ppm and tried inferencing but it could not infer properly.

I need at least 10000 such .ppm images for experimenting. Let me know from which source the ppm file for the image airliner.ppm you have taken
Otherwise let me know a procedure how to convert .jpg or .jpeg images to .ppm format

Apart from this please update me with all the queries I have raised.

Thanks and Regards
Nagaraj Trivedi

SivaRamaKrishnaNV · January 9, 2024, 3:55pm

Dear @trivedi.nagaraj,

The sampleINTAPI already uses dynamic range from /usr/src/tensorrt/data/int8_api /resnet50_per_tensor_dynamic_range.txt and prepares INT8 model.
If you want to perform calibration using a set of images(use atleast 500 images) and prepare calibration file. You may find calibrator implementation at /usr/src/tensorrt/samples/ommon/sampleEngines.cpp . As you are on different Jetapack release please check grep -lir "writeCalibrationCache" to find calibrator implementation.
Currently, The ResNet50 model is static model with batchsize 1
You can use any dataset like imagenet to get input images for your test. You can use convert linux tool to convert jpeg → ppm format for your test.

trivedi.nagaraj · January 9, 2024, 5:56pm

Hi ShivaRamaKrishna, thank you for your response clarifying my doubts.
Please find my comments below your points mentioned for 1 to 4

Yes there is a file called resnet50_per_tensor_dynamic_range.txt but wanted to get it confirmed with you
I will look into the sampleEngines.cpp to know how the calibration file is created. I have already looked at it before and look it again
How I can modify it to work with different batch sizes, please clarify.
I have read this in the tensorrt document

If the input model is in ONNX format, use the --minShapes, --optShapes, and --maxShapes flags to control the range of input shapes including batch size.
*** --minShapes=, --optShapes=, and --maxShapes=: Specify the range of the input shapes to build the engine with. Only required if the input model is in ONNX format.**
I believe when we use these options through command line then it might be internally performing some operation that will result into building or inferencing the engine with different batch size

I tried converting jpg → ppm but the prediction was wrong. Is that I need to resize and normalize and perform transpose of jpg or jpeg image before converting to the ppm

Please clarify all these doubts.

Thanks and Regards

Nagaraj Trivedi

trivedi.nagaraj · January 10, 2024, 5:42am

In continuation to this, I tried inferencing with a jpg image it thrown error. It is due the reason that in the function prepareInput() it checks the file format to in .ppm. I even tried commenting it and it could infer but falsely. I request you to once verify inferring with either .jpg or .jpeg image
Please see the below code
bool SampleINT8API::prepareInput(const samplesCommon::BufferManager& buffers)
{
sample::gLogInfo<<"Image file name is "<<mParams.imageFileName<<std::endl;
if (samplesCommon::toLower(samplesCommon::getFileType(mParams.imageFileName)).compare(“ppm”) != 0)
** {**
** sample::gLogError << “Wrong format: " << mParams.imageFileName << " is not a ppm file.” << std::endl;**
** return false;**
** }**

Thanks and Regards

Nagaraj Trivedi

AastaLLL · January 10, 2024, 7:34am

Hi,

3. How I can modify it to work with different batch sizes, …

Since the model is static, please modify the model parameter directly.
A sample can be found in the below comment:

4. I tried converting jpg → ppm but the prediction was wrong. …

The image needs to be resized into 3x224x224 (CHW).
But the normalization and scaling are done in cod already:

github.com

NVIDIA/TensorRT/blob/release/8.2/samples/sampleINT8API/sampleINT8API.cpp#L444


      
                  for (int h = 0; h < height; ++h)
                  {
                      for (int w = 0; w < width; ++w)
                      {
                          int dstIdx = c * height * width + h * width + w;
                          int srcIdx = h * width * channels + w * channels + c;
                          // This equation include 3 steps
                          // 1. Scale Image to range [0.f, 1.0f]
                          // 2. Normalize Image using per channel Mean and per channel Standard Deviation
                          // 3. Shuffle HWC to CHW form
                          hostInputBuffer[dstIdx] = (2.0 / 255.0) * static_cast<float>(fileData[srcIdx]) - 1.0;
                      }
                  }
              }
              return true;
          }
          
          //!
          //! \brief Verifies that the output is correct and prints it
          //!
          bool SampleINT8API::verifyOutput(const samplesCommon::BufferManager& buffers) const

Thanks.

trivedi.nagaraj · January 10, 2024, 7:49am

Hi, thank you for your response. I tried searching in the internet, but not finding a proper website to download different images which are of format .ppm
If you have multiple such .ppm images then please provide me. Otherwise let me know a way how can I convert the .jpg or .jpeg to the .ppm format.

Thanks and Regards

Nagaraj Trivedi

trivedi.nagaraj · January 10, 2024, 8:28am

trivedi.nagaraj:

trivedi.nagaraj:

I tried converting jpg → ppm but the prediction was wrong. Is that I need to resize and normalize and perform transpose of jpg or jpeg image before converting to the ppm

In continuation to this, I tried inferencing with a jpg image it thrown error. It is due the reason that in the function prepareInput() it checks the file format to in .ppm. I even tried commenting it and it could infer but falsely. I request you to once verify inferring with either .jpg or .jpeg image
Please see the below code
bool SampleINT8API::prepareInput(const samplesCommon::BufferManager& buffers)
{
sample::gLogInfo<<"Image file name is "<<mParams.imageFileName<<std::endl;
if (samplesCommon::toLower(samplesCommon::getFileType(mParams.imageFileName)).compare(“ppm”) != 0)
** {**
** sample::gLogError << “Wrong format: " << mParams.imageFileName << " is not a ppm file.” << std::endl;**
** return false;**
** }**

Thanks and Regards

Nagaraj Trivedi

Hi, please address the issue also which I have asked. The reason is there is only one image airliner.ppm file but I need around 10000 images. I tried searching in the internet but could not find them.
This sample takes only .ppm file and doesn’t accept others. I tried commenting the code that checks for .ppm format and fed the jpg image but it could not predict it properly.
If you have collection of .ppm files or know the link from where I can download them then let me know. Otherwise let me know how to modify this code to work with both jpg and jpeg format or a way how I can convert jpg or jpeg files to the ppm format

Here is the code that checks for the ppm format in the sample code

In continuation to this, I tried inferencing with a jpg image it thrown error. It is due the reason that in the function prepareInput() it checks the file format to in .ppm. I even tried commenting it and it could infer but falsely. I request you to once verify inferring with either .jpg or .jpeg image
Please see the below code
bool SampleINT8API::prepareInput(const samplesCommon::BufferManager& buffers)
{
sample::gLogInfo<<"Image file name is "<<mParams.imageFileName<<std::endl;
if (samplesCommon::toLower(samplesCommon::getFileType(mParams.imageFileName)).compare(“ppm”) != 0)
** {**
** sample::gLogError << “Wrong format: " << mParams.imageFileName << " is not a ppm file.” << std::endl;**
** return false;**
** }**

Thanks and Regards

Nagaraj Trivedi

SivaRamaKrishnaNV · January 11, 2024, 8:54am

Dear @trivedi.nagaraj ,
ImageNet dataset : ImageNet
PASCAL VOC dataset : The PASCAL Visual Object Classes Challenge 2012 (VOC2012)

Convert utility : https://linux.die.net/man/1/convert
somr stackoverflow reference : jpeg - How to convert images from the jpg format to ppm(P3) - Stack Overflow
command line - How to resize an image through the terminal? - Ask Ubuntu

SivaRamaKrishnaNV · January 16, 2024, 3:30am

You may use coco dataset(COCO - Common Objects in Context) as well.

Topic		Replies	Views
With INT8 the accuracy of the network_api_pytorch_mnist model drops to less than 10 % Jetson Orin Nano pytorch	24	871	January 2, 2024
TensorRT fails to build FasterRCNN GIE model with using INT8 TensorRT	28	9206	May 3, 2018
X264 and TensorRT sudden reboot (MJPG encoder not affected, but not fast enough) on Jetson Orin Nano Jetson Orin Nano tensorrt , jetson-inference , gstreamer , jetson	52	798	June 18, 2024
ONNX model with Jetson-Inference using GPU Jetson Xavier NX tensorrt , jetson-inference , onnx	38	5603	October 18, 2021
Calibration and int8 inference on Onnx model TensorRT tensorrt	17	2364	March 20, 2023
Batch processing on tao engine TAO Toolkit	11	397	March 27, 2024
Error generated while running the code after connecting the camera Jetson Xavier NX gstreamer , nvbugs	45	1233	January 2, 2024
[ERROR] Model has dynamic shape but no optimization profile specified. Aborted (core dumped) TAO Toolkit	30	2020	December 13, 2021
TensorRT INT8 inference is slower than FP16 in models with conditional flow Jetson Orin Nano tensorrt , cuda , jetson-inference , onnx	5	831	June 10, 2024
Get wrong result when I using tensorRT to do inference, am I wrong to use ? Jetson TX2	18	2985	October 18, 2021

Code modified for CUDA Graph in the sampleINTAPI is throwing error

Related topics