rogerallen/raytracinginoneweekendincuda

ptxas' died due to "signal 11" (Invalid memory reference)

josiahblaisdell opened this issue · 2 comments

Hi,

When I attempted to write my own ray tracer using this code as a guide I encountered an error where ptxas would crash.

Unfortunately it is not reproducing now, I encountered the bug about a month ago. I have rewritten a lot of my code and I have not been using source control. This is embarrassing, it was so easy to reproduce just a month ago.

I went through my google search history and the error I was encountering was:
"ptxas' died due to "signal 11" (Invalid memory reference)"
"stack size for entry function cannot be statically determined"

Here is a similar error that I found when I was debugging it:
https://devtalk.nvidia.com/default/topic/808186/nvcc-error-ptxas-died-due-to-signal-11-invalid-memory-reference-/

It seemed to be caused by the dielectric material's scatter function. The scatter function you have is:
__device__ virtual bool scatter(const Ray& r_in, const HitRecord& rec, vec3& attenuation, Ray& scattered, curandState *local_rand_state) const = 0;

The scatter function I wrote was:
__device__ virtual bool scatter(const Ray& r_in, const HitRecord& rec, vec3& attenuation, Ray& scattered, curandState *local_rand_state, bool rand) const = 0;

Where "rand" is used in the dielectric function as:
rand < reflect_prob ? scattered = Ray(rec.p, reflected) : scattered = Ray(rec.p, refracted);

I am not sure why rand is a bool and not a float. It has been awhile since I touched this code but this is exactly what I did to work around this problem. I do have a record of that in a comment.

Here are the details on the system I am using:
image

I was able to debug this by looking at the ptxas output. To do this I use the following compile command:
nvcc -ccbin g++ -I. -I/usr/local/apps/cuda/cuda-9.2/include -gencode arch=compute_35,code=sm_35 --ptxas-options="-v" -c $< -o $@ --verbose

Since this is happening with derived code, it seems best to just close this out. Others may find this with a google search & perhaps it can be helpful in some way. Thanks for the report.

Thanks!