|
@@ -7,7 +7,10 @@
|
|
|
#include "gdrapi.h"
|
|
|
#include <pcilib.h>
|
|
|
#include <pcilib/kmem.h>
|
|
|
+#include <pcilib/bar.h>
|
|
|
#include "ipedma.h"
|
|
|
+#include <unistd.h>
|
|
|
+
|
|
|
|
|
|
int main()
|
|
|
{
|
|
@@ -43,6 +46,9 @@ int main()
|
|
|
assert_cu( cuCtxCreate(&cuCtx,0,GPU) );
|
|
|
assert_cu( cuCtxGetDevice(&GPU) );
|
|
|
printf("Device for this context: %d\n",GPU);
|
|
|
+ int pi;
|
|
|
+ assert_cu( cuDeviceGetAttribute(&pi,CU_DEVICE_ATTRIBUTE_UNIFIED_ADDRESSING,GPU) );
|
|
|
+ printf("Support unified addressing? %d\n",pi);
|
|
|
|
|
|
/* Allocate memory on the device, pin and map */
|
|
|
printf("\nMemory mapping with the GPU\n");
|
|
@@ -83,9 +89,12 @@ int main()
|
|
|
}
|
|
|
|
|
|
CUdeviceptr dBAR;
|
|
|
+ int getdBAR = 10;
|
|
|
assert_cu( cuMemHostRegister((void*)pciVa,128,CU_MEMHOSTREGISTER_IOMEMORY) );
|
|
|
assert_cu( cuMemHostGetDevicePointer(&dBAR,(void*)pciVa, 0) );
|
|
|
printf("pciVa = %p\ndBar = 0x%llx\n",pciVa,dBAR);
|
|
|
+ assert_cu( cuMemcpyDtoH(&getdBAR,dBAR,sizeof(int)) );
|
|
|
+ printf("getdBAR = %d\n",getdBAR);
|
|
|
|
|
|
pcilib_kmem_handle_t* pciHandle;
|
|
|
pciHandle = pcilib_alloc_kernel_memory(pciCtx,PCILIB_KMEM_TYPE_CONSISTENT,1,64,4096,PCILIB_KMEM_USE_STANDARD,PCILIB_KMEM_FLAG_REUSE);
|
|
@@ -109,26 +118,55 @@ int main()
|
|
|
exit( EXIT_FAILURE );
|
|
|
}
|
|
|
|
|
|
- printf("pciMemPtr = %p\npciBus = %p or %lu\n",pciMemPtr,pciBus,pciBus);
|
|
|
+ printf("pciMemPtr = %p\npciBus = %p\n",pciMemPtr,pciBus);
|
|
|
+
|
|
|
+ const pcilib_bar_info_t* bar_info;
|
|
|
+ bar_info = pcilib_get_bar_info(pciCtx,0);
|
|
|
+ if( bar_info == NULL )
|
|
|
+ {
|
|
|
+ printf("Cannot get BAR info\n");
|
|
|
+ exit( EXIT_FAILURE );
|
|
|
+ }
|
|
|
+ else
|
|
|
+ printf("bar = %d\nsize = %lu\nphys_addr = %p\nvirt_addr = %p\n",bar_info->bar,bar_info->size,bar_info->phys_addr,bar_info->virt_addr);
|
|
|
|
|
|
/* Try some stuff... */
|
|
|
printf("\nRunning some tests\n");
|
|
|
*(int*) pciMemPtr = 65;
|
|
|
printf("Try to read pciMem in another way : %d %d\n",*(int*)pciMemPtr,*(int*)dBAR);
|
|
|
assert_cu( cuMemcpyHtoD(dptr,(void*)pciMemPtr,sizeof(int)) );
|
|
|
+ assert_cu( cuMemcpyHtoD(dBAR,(void*)pciMemPtr,sizeof(int)) );
|
|
|
+ assert_cu( cuMemcpyDtoH(&getdBAR,dBAR,sizeof(int)) );
|
|
|
+ printf("getdBAR = %d\n",getdBAR);
|
|
|
printf("Try to read pciMem in another way : %d %d\n",*(int*)pciMemPtr,*(int*)dBAR);
|
|
|
printf("Try to read pciMem in another way : %d\n",*(int*)va);
|
|
|
printf("A small computation\n");
|
|
|
add_three_global<<< 1,1 >>>(dptr);
|
|
|
- cudaDeviceSynchronize();
|
|
|
+ assert_cuda( cudaDeviceSynchronize() );
|
|
|
int foo;
|
|
|
assert_gdr( gdr_copy_from_bar(&foo,va,sizeof(foo)) );
|
|
|
printf("foo = %d\nva = %d\n",foo,*(int*) va);
|
|
|
+ printf("Now the other way around\n");
|
|
|
+ foo = 100;
|
|
|
+ assert_gdr( gdr_copy_to_bar(va,&foo,sizeof(foo)) );
|
|
|
+ add_three_global<<< 1,1 >>>(dptr);
|
|
|
+ assert_cuda( cudaDeviceSynchronize() );
|
|
|
+ assert_cu( cuMemcpyDtoH((void*)pciMemPtr,dptr,sizeof(int)) );
|
|
|
+ printf("pciMem = %d\n",*(int*)pciMemPtr);
|
|
|
+
|
|
|
+ /* FPGA */
|
|
|
+ printf("\nWorking on the FPGA\n");
|
|
|
+ WR32 (REG_RESET_DMA, 1);
|
|
|
+ usleep (100000);
|
|
|
+ WR32 (REG_RESET_DMA, 0);
|
|
|
+ usleep (100000);
|
|
|
+
|
|
|
+
|
|
|
|
|
|
/* Close everything */
|
|
|
+ printf("\nClosing the connections\n");
|
|
|
assert_cu( cuMemHostUnregister((void*) pciVa) );
|
|
|
pcilib_close(pciCtx);
|
|
|
- printf("\nClosing the connections\n");
|
|
|
assert_gdr( gdr_unmap(g,GPUMemHandle,va,(size_t) GPUProp.sharedMemPerBlock) );
|
|
|
assert_gdr( gdr_unpin_buffer(g,GPUMemHandle) );
|
|
|
assert_gdr( gdr_close(g) );
|