Hacker News Board Hacker News, browsed like a classic message board
Live data from Hacker News

What happens when you run a CUDA kernel?

fergusfinn.com

Post reply on HN
Print view View original

21–30 of 36 posts

Re: What happens when you run a CUDA kernel?

#21

Post by spmurrayzzz »

xpct wrote: post #18
spmurrayzzz wrote: post #7

Earlier quoted context omitted.

Near-term acquihires are certainly a likely bet I think. But given model progress on related benchmarks like kernelbench [1], I do think a set of more commoditized solutions is also inevitable. The caveat though is that each new gen of hardware often comes with brand new constraints/features that a given generation of models haven't seen before (e.g. tcgen05 in blackwell was OOD at one point). As the models start to…

Thank you for sharing the link. It's fascinating that the models can only do 10-20% on the hard subset, and I wonder why that is so. The fact that they can only get 30-40% out of the fp8 GEMM seems unintuitive to me, I would've expected a convergence near ~80%.

I'm not entirely up to date with the latest batch, but I've reviewed some of the rollouts in the past and my sense is that the models are surprisingly good at getting correct custom kernels in the happy path, but still weak at sustained/shape-robust workloads. Having to deal with writing the full path from scratch compounded by weird memory layouts, odd sizes, routing, unpacking quantized weights, etc. is definitely challenging.

Also, at least a portion of this you could argue is arbitrary and entirely scoped to the eval itself. The fp8 GEMM score could be low simply because one of the shapes is fairly skinny (i.e. not enough math work to keep the compute engine busy for a meaningful amount of time).

Re: What happens when you run a CUDA kernel?

#22

Post by saagarjha »

connicpu wrote: post #11
orliesaurus wrote: post #5

There are companies whose whole job right now is to optimize kernels so that things run faster. I wonder if those companies are going to be dethroned by some sort of like open source library that can do that really well (I bet Nvidia could release it any day.).. or if they're going to thrive and be acquired by the big providers as a `moat` to speed up their infrerence.

When you run CUDA at scale dealing with nvidia driver and library bugs takes up a disgustingly large percentage of engineer time, I don't know a lot of people who would be looking forward to rely on more nvidia libraries.

The alternative is you spend even more time on a competitor's system dealing with their bugs

Re: What happens when you run a CUDA kernel?

#23

Post by saagarjha »

orliesaurus wrote: post #15
einpoklum wrote: post #13

Earlier quoted context omitted.

Probably not, because the specifics of the workload - exact parameters, representation of data in memory, value ranges etc - lead you to highly divergent optimization strategies.

shouldn't it be possible to be run as a mlautoresearch project? i.e. orchestrate 10 strategies to speed it up, run in paralellel, pick the winning and go from there?

No, because all the low hanging fruit that this kind of thing would find has usually been picked.

Re: What happens when you run a CUDA kernel?

#24

Post by aberrahmane_b »

It's very useful. The doorbell and QMD part were the most useful for me, because it connects the CUDA launch syntax to what actually gets submitted to the GPU. Most explanations stop around kernels, blocks and warps, but this made the CPU to driver to GPU path much easier to follow.

Re: What happens when you run a CUDA kernel?

#26

Post by Athas »

david-gpu wrote: post #19
connicpu wrote: post #11

Earlier quoted context omitted.

When you run CUDA at scale dealing with nvidia driver and library bugs takes up a disgustingly large percentage of engineer time, I don't know a lot of people who would be looking forward to rely on more nvidia libraries.

How do you determine that the bugs you run into are located in the Nvidia drivers and libraries? Way back when I wrote the OpenCL driver at Qualcomm, we would frequently get bug reports from customers complaining about our code. During my tenure, every single one of them was root-caused as an application bug. Unsurprisingly, considering that our code was backed by an extensive test suite and their code wasn't. Not to…

I have never used Qualcomm's OpenCL driver, but it is not unknown to get the NVIDIA driver into a state where some kernel is stuck in a running state, or some memory is allocated long after the originating process has terminated. This is usually down to application bugs, sure - but no application bug should be able to wedge the driver. While developing GPU kernels, the code will certainly be buggy, and hence the driver should be robust. For that matter, maybe I am running untrusted GPU code, and anytime the driver gets in a weird or stuck state, I am uneasy that it might not be many steps away from an exploitable situation. We don't accept this in CPU operating systems, so why should it be acceptable for GPUs? We are talking unprivileged code - nothing runs as root. Ever since I first got into GPGPU programming (about 2012), I noticed that they were far less robust in the face of buggy code than I was accustomed to.

It is also common in my experience for buggy GPU code to crash displays if the GPU is simultaneously used to drive a monitor. This usually happens for kernels that go into infinite loops, or out-of-memory conditions.

It is my understanding that modern GPU drivers even have watchdog systems that notice when they get stuck and forcibly reboot them, which to me is mere symptom treatment.

Re: What happens when you run a CUDA kernel?

#28

Post by keynha »

orliesaurus wrote: post #5

There are companies whose whole job right now is to optimize kernels so that things run faster. I wonder if those companies are going to be dethroned by some sort of like open source library that can do that really well (I bet Nvidia could release it any day.).. or if they're going to thrive and be acquired by the big providers as a `moat` to speed up their infrerence.

[flagged]

Re: What happens when you run a CUDA kernel?

#29

Post by AlotOfReading »

connicpu wrote: post #20
david-gpu wrote: post #19

Earlier quoted context omitted.

How do you determine that the bugs you run into are located in the Nvidia drivers and libraries? Way back when I wrote the OpenCL driver at Qualcomm, we would frequently get bug reports from customers complaining about our code. During my tenure, every single one of them was root-caused as an application bug. Unsurprisingly, considering that our code was backed by an extensive test suite and their code wasn't. Not to…

If you're big enough you can get direct access to Nvidia engineers, and they are usually transparent when they find out the bug was in their software and send you a patched version to try to resolve the issue

And when the bug is in hardware instead, good luck getting them to admit it. Instead they go "We can't confirm that there's an issue, but it'll be gone in the next revision and meanwhile don't use that instruction sequence".

Re: What happens when you run a CUDA kernel?

#30

Post by david-gpu »

AlotOfReading wrote: post #29
connicpu wrote: post #20

Earlier quoted context omitted.

If you're big enough you can get direct access to Nvidia engineers, and they are usually transparent when they find out the bug was in their software and send you a patched version to try to resolve the issue

And when the bug is in hardware instead, good luck getting them to admit it. Instead they go "We can't confirm that there's an issue, but it'll be gone in the next revision and meanwhile don't use that instruction sequence".

That is largely what drivers do: work around hardware bugs. It's the industry's dirty secret.

To be fair: the hardware is enormously complex, and the drivers much less so.

Post reply on HN