Access Violation while launching Graph

I have a series of async operations I’ve captured in a graph. I wrapped them all in a function for easy writing, so graph capture looks like:

cudaStreamBeginCapture(primaryStream, cudaStreamCaptureModeGlobal);

		// run kernels/events for the stream capture
		binningLaunch();

		cudaStreamEndCapture(primaryStream, &inGraph);

		auto err = cudaGraphInstantiate(&ingestGraph, inGraph, NULL, NULL, 0);
		cudaErrCheck(err, "Graph Create Fail");

primaryStream, inGraph and ingestGraph are all members of the object this code is being called from.

Later on i call this code with cudaGraphLaunch(ingestGraph, primaryStream)

and I get:

Access violation reading location 0xFFFFFFFFFFFFFFFF.

And program crashes.

When I call binningLaunch() normally, everything works fine and the kernels go do their thing. Memcheck is clean, no launch errors or anything. No errors from the graph creation either.

So what might be going on inside the Graph that is causing an access violation?

-del-

My only guess is that the pointer references in the for loop aren’t being captures as I intend and the graph is looking for a reference that doesn’t exist. But I would think that when the captured kernels are launched, the dereference means that the stream capture is effectively seeing a value that in then copies.

Not sure what is going on here. Any insights or pointers appreciated.

Turns out I was changing the address in the pointers I was using as src for the mem copies after I made the graph. Once I fixed that by keeping the src constant, no more graph failure.

Removing some of my code from earlier for reasons.

This topic was automatically closed 14 days after the last reply. New replies are no longer allowed.