Weird behaviour of device function parameter - "Variable is not live at this PC."

Hi all,

This one has had me stumped for a few hours, I’m trying to pass a pointer to a float3 into the following function:

static __inline__ __host__ __device__ float3 LocalizePosition(const float3 *globalPosition, const vehicle_data *vehicleData)

	{

		// Global offset from local origin

		float3 globalOffset = float3_subtract(*globalPosition, vehicleData->position);

		return LocalizeDirection(&globalOffset, vehicleData);

	}

calling code…

const float3 *center = &obstacleData[obstacleIndices[i]].center;

			const float radius = obstacleData[obstacleIndices[i]].radius;

			// Transform the obstacle into local space

			float3 localPosition = LocalizePosition(center, &VDATA(offset));

complete source file: [attachment=18676:AvoidObs…DAKernel.cu]

Watching the variable using Nsight outside of the function, it is set and valid, but inside of the function the debugger reports “Variable is not live at this PC.”, and further calculations on it use the value (0,0,0) instead of what should be there. Searching on forums and Google turn up a single similar case which is unresolved.

I’m using 32bit 3.2 toolkit/SDK, Nsight 1.5 (64bit) on 64bit Win7.

Does anyone have an idea why this would happen or a workaround? I’m happy to provide more code if required.

Cheers,

Owen
AvoidObstaclesCUDAKernel.cu (4.59 KB)

Hi all,

This one has had me stumped for a few hours, I’m trying to pass a pointer to a float3 into the following function:

static __inline__ __host__ __device__ float3 LocalizePosition(const float3 *globalPosition, const vehicle_data *vehicleData)

	{

		// Global offset from local origin

		float3 globalOffset = float3_subtract(*globalPosition, vehicleData->position);

		return LocalizeDirection(&globalOffset, vehicleData);

	}

calling code…

const float3 *center = &obstacleData[obstacleIndices[i]].center;

			const float radius = obstacleData[obstacleIndices[i]].radius;

			// Transform the obstacle into local space

			float3 localPosition = LocalizePosition(center, &VDATA(offset));

complete source file: [attachment=24223:AvoidObs…DAKernel.cu]

Watching the variable using Nsight outside of the function, it is set and valid, but inside of the function the debugger reports “Variable is not live at this PC.”, and further calculations on it use the value (0,0,0) instead of what should be there. Searching on forums and Google turn up a single similar case which is unresolved.

I’m using 32bit 3.2 toolkit/SDK, Nsight 1.5 (64bit) on 64bit Win7.

Does anyone have an idea why this would happen or a workaround? I’m happy to provide more code if required.

Cheers,

Owen

Since you provided the .cu file, could you please also attach the header files you are using?
Otherwise I cannot compile it…

Since you provided the .cu file, could you please also attach the header files you are using?
Otherwise I cannot compile it…

There’s a whole lot of headers and dependencies so its probably easier if I attach the whole source for the project and the test project. The kernel code has been modified slightly in this version but the same problem still persists.

Update: new code, glut included and include/lib paths fixed. Should compile as is now.

opensteer.zip (351 KB)

There’s a whole lot of headers and dependencies so its probably easier if I attach the whole source for the project and the test project. The kernel code has been modified slightly in this version but the same problem still persists.

Update: new code, glut included and include/lib paths fixed. Should compile as is now.
[attachment=24346:opensteer.zip]

Same problem with Nsight 1.0/3.1 toolkit, on both GTS250 and GTX285. I’ve gotta be doing something dumb but I still can’t see it.

Same problem with Nsight 1.0/3.1 toolkit, on both GTS250 and GTX285. I’ve gotta be doing something dumb but I still can’t see it.

What happens if you dump the const.

What happens if you dump the const.

Same thing without the const. I thought maybe it was because I was passing a pointer to local memory into the device function which was expecting a pointer to global memory, so I modified it again to pass the global memory pointer

&obstacleData[obstacleIndices[i]].center

directly into the function. The value seems to be fine until I try and use it, then the “Variable is not live at this PC.” error occurs and it gets treated as a float3 with value 0,0,0. I’ve uploaded a short clip to YouTube here showing what is happening in the debugger. Its still processing on the server but should be available (and hopefully readable) soon - if not I’ll fix it up when I get home later.

Same thing without the const. I thought maybe it was because I was passing a pointer to local memory into the device function which was expecting a pointer to global memory, so I modified it again to pass the global memory pointer

&obstacleData[obstacleIndices[i]].center

directly into the function. The value seems to be fine until I try and use it, then the “Variable is not live at this PC.” error occurs and it gets treated as a float3 with value 0,0,0. I’ve uploaded a short clip to YouTube here showing what is happening in the debugger. Its still processing on the server but should be available (and hopefully readable) soon - if not I’ll fix it up when I get home later.

As YouTube is still “processing” the video I uploaded around 6 hours ago, I’ll post an avi (xvid) here as well.

20101103_1223_33_xvid.avi (2.44 MB)

As YouTube is still “processing” the video I uploaded around 6 hours ago, I’ll post an avi (xvid) here as well.
[attachment=24314:20101103_1223_33_xvid.avi]