Cuda encoder

Hi
My aim is to do H264 video encoding on GPU on a Linux platform.
For few days, I looking for an already implemented library, watching conferences on gpu video encoding and I found CUDA video Encoder. ( http://developer.download.nvidia.com/compute/DevZone/docs/html/C/doc/CUDA_VideoEncoder_Library.pdf )
It looks like so usefull i didn’ found any linux version.
Did I miss it? if not, why is there no linux version? Is there so other free alternatives?

Thanks

Hi

I have also the task to encode raw YUV2 video frames with CUDA on Debian Linux. I have installed CUDA GPU Computing SDK 4.1 but there is a difference between this SDK on Windows and Debian. On Windows C sources contains cudaEncode sample which is missing from Linux version.
Could someone help where can I find and download the libraries for H.264 encoding on Debian?

Many thanks,
Arpad

Hi,
look here: CUDA Samples :: CUDA Toolkit Documentation

it is available only for Windows.

But I also want to know why there is no version for Linux?

Thanks

so guys… did any of you got it in linux

For now I still have no news about the nvidia encoder.
From the x264 mailing list, i get a h264 video encoder based on cuda, but it supports only the full hd resolution. I’ll try to make it works with others resolution.
here is the link with source code:
http://masa.nudt.edu.cn/h264.html

If someone has any information about the official nvidia video encoder…

Hi, ive try to use that implementation on Ubuntu 12.04 32 bits, but i cant, its tell me that: error: calling a host funtion (“std::abs”) from a device/global funtion (“cavlc_texture_symbols_chroma_DC_kernel”) is not allowed.

I see there is not an abs funtion on that kernel, and i dont know how to include it, can you tell me how to do that?

Hi.

I didn’t get the problem with CUDA SDK 4.0 on a linux (ubuntu 10.10, maverick).

Every thing was working at full HD resolution (1920*1080)

If you still can’t compile it, try to replace this function by your own device abs implementation.

ahh, i dont do that because i believe it’s special funtion, but thanks, i going to try both solutions

hi again xD, emm i did it, but now i have other problem, when i try to compress videos within AVI format, i seen that the format the codec compresses is YUV, what others formats can i use?, and, when you say the codec only compresses HD videos, its 720p and 1080p videos?, and, how long is the maximum duration of the compressed video?, because i tried with a video of 1:30 hrs

By default, the main function read from a file in which the video is in YUV (YV12) format. Many encoders use YUV as input because YUV is a format in which the “only” important channel is Y because the 2 others only have informations about luminosity and colours so they can be more compressed ( Recommended 8-Bit YUV Formats for Video Rendering - Win32 apps | Microsoft Learn )

If you want to convert a video from an other colorspace than YUV you will have to transcode them in YUV first. I did a cuda kernel which is doing that (RGB=>YV12) if you’re interested.

I think initially the encoder support only 180p videos. I did some modification in the code to make it able to encode others resolutions too. But there are no time limits. If there is one, it is a bug (memory leak probably) .

To sum up, the encoder can only take YUV videos as input to create h264 rawstreams. If you need tu use it with an other input you will have yo code something to convert your initial video in YUV and if you want to encapsulate the generated h264 stream in a container (such as MP4, avi…) you will have to do it yourself.

Hello,

i have been following this thread. its good to know that you did cuda kernel for (RGB=>YV12), can you share that with us?

Thanks,

Amature

Of course I can.

In attachments there are:

-cudaPixel.cuh some usefull function about pixel processing

-rgb2Yv12Kernel.cu the transcoding kernel

-writeYUVframe.cu a c file which write some yuv files to validate them with common players

You can use the kernel with

// define one block of X_NB_SLICES * Y_NB_SLICES threads

dim3 threads(X_NB_SLICES, Y_NB_SLICES);

dim3 block(1,1);

rgb2Yv12Kernel <<<block,threads>>> ( . . . )

Let’s precise that the input format is RGBA so 32bits/ pixels

I hope it doesn’t miss any file, Good luck
cudaPixel.cu (3.22 KB)
rgb2Yv12Kernel.cu (6.71 KB)

thank you very much my friend, you dont know how much you help me, Im a developer with little experience in this field, and Im studying a little of everything

You’re welcome.

I obtained interesting performances with it but I think it could be optimizeda little bit more. I’m trying to split the image in 16x16 macroblocks as in the h264 norme but for now I didn’t succeded with it.

Can you suggest me a YUV to AVI converter, or other formats, for ubuntu of course, I have one for Windows, but I can not use it for my purposes.

Avi is a container. It means that you can have different video format in it. Here we are talking about a h264 encoder. So if you want an avi output file you “just” have to wrap it in an AVI file. But this is not done by this encoder, you will have to implement it. I don’t know any cuda encoder for others video format so I can’t help you more sorry.

@xterwolf error: calling a host funtion (“std::abs”) from a device/global funtion (“cavlc_texture_symbols_chroma_DC_kernel”) is not allowed.

how did u solved this error…
hoping to get reply soon!

I do not know what version of ubuntu you’re using, I use the 12.04, the error came because this version of ubuntu is not supported by CUDA yet, so I install ubuntu 10.10

OK…well, i was trying to make this code run on Fedora16 not ubuntu…anyways thanks for the quick reply…

the error is solved now …i made my own function for std::abs() and now its working f9.

Thanks!

has any one able to encode video of 720p in this code…currently, it encodes only 1080p yuv file.
i tried to change the frame width and height accordingly but the output obtained cannot be decoded by the decoder.

any clue regarding this…am new to this video coding.

Thanks