One-byte offset from copy to __constant__

Update: NEVER MIND. I forgot the ‘offset’ parameter in my memcpy.

Here’s some pseudocode:

void host calling_func() {
int i;
my_struct s[FIXED_SIZE];
for (i = 0; i < FIXED_SIZE; i++)
assign_stuff_to_my_struct(&s[i]);
cudaMemcpyToSymbol( “s”, s, sizeof(my_struct) * FIXED_SIZE, cudaMemcpyHostToDevice);
kernel_func<<<1, 1>>>();
}

The kernel’s .cu has a

constant my_struct s[FIXED_SIZE];

at file scope. When I’m in the kernel, using deviceemu and gdb, every value is shifted one byte to the left. For example, a pointer stored in my_struct in host code has a value of 0x157ae00, while on the device, it’s 0x157ae0000, and it happens in the same manner with ints, shorts, and floats. (It’s not an arithmetic shift, it’s an address window shift; a short with a value of 0x0008 (8) on host becomes 0x083f (2111) on device.) sizeof(my_struct) is the same on both architectures. Any hints as to why this is happening?