Starting CUDA profiling: initialization error #395
Comments
|
it is ok with this: git diff .
diff --git a/examples/other/vectorAdd_profiled.cu b/examples/other/vectorAdd_profiled.cu
index a31c937..22408a1 100644
--- a/examples/other/vectorAdd_profiled.cu
+++ b/examples/other/vectorAdd_profiled.cu
@@ -25,8 +25,23 @@ __global__ void vectorAdd(const float *A, const float *B, float *C, int numEleme
if (i < numElements) { C[i] = A[i] + B[i]; }
}
+void initCuda()
+{
+ CUresult a;
+ CUcontext pctx;
+ CUdevice device;
+ cuInit(0);
+ cuDeviceGet(&device, 0);
+ a = cuCtxCreate(&pctx, 0, device );
+ assert(a == CUDA_SUCCESS);
+ std::cout << "Initialized CUDA" << std::endl;
+}
+
int main()
{
+ auto res = cuProfilerInitialize ( "prof.conf", "profiler.csv", CU_OUT_CSV );
+
+ initCuda();
profile_this_scope();
cuda::profiling::name_this_thread("The single thread for vectorAdd_profile :-)");
if (cuda::device::count() == 0) {as |
|
Acknowledged. This is a regression I somehow missed with making sure everything is covered by the appropriate driver-related initializations. Will work on this hopefully today and otherwise later this week. Sorry about this. |
… profiling * `context::current::detail_::scoped_existence_ensurer_t` will now initialize the CUDA driver if necessary - as part of creating a context when none exists. * The profiling `scope` class now has a `scoped_existence_ensurer_t` member - which ensures the driver is initialized and that some context is current when profiling begins.
|
Please try the HEAD of the development branch, which now has the fix. |
|
Great! Thank you @eyalroz , it would be nice to enrich example to get some profile stats, e.g. ipc, memory bandwidth as the example belongs to BLAS level I, it is memory bound. |
|
Hi @eyalroz , just checked from [alsam@Noire build_debug2]$ cmake .. -DCMAKE_BUILD_TYPE=Debug -DCAW_BUILD_EXAMPLES=ON
...
cd exampes/bin
$ ./vectorAdd_profiled
terminate called after throwing an instance of 'cuda::runtime_error'
what(): Starting CUDA profiling: initialization error |
|
@alsam : Are you sure you checked out from the develop branch? |
|
@eyalroz maybe I missed something? From which branch should I get the code? See your changes in |
|
@alsam :Yes, you missed the branch change. You build the master branch - and fixes don't go there, they go on switch the branch by executing |
|
ok, it works! git checkout development
...
./vectorAdd_profiled
CUDA kernel launch with 1954 blocks of 256 threads
SUCCESSThanks! |
… profiling * `context::current::detail_::scoped_existence_ensurer_t` will now initialize the CUDA driver if necessary - as part of creating a context when none exists. * The profiling `scope` class now has a `scoped_existence_ensurer_t` member - which ensures the driver is initialized and that some context is current when profiling begins.
|
Thank you for taking the time to report this and checking the fix. If end up using the library for an interesting project, or even if you just feel like giving some library usability feedback, do drop me a line... |
|
Thank you for the great library @eyalroz ! Envisage a potential usage for e.g. collecting traces for different workloads, replaying them, collecting some of performance counters :-) |
… profiling * `context::current::detail_::scoped_existence_ensurer_t` will now initialize the CUDA driver if necessary - as part of creating a context when none exists. * The profiling `scope` class now has a `scoped_existence_ensurer_t` member - which ensures the driver is initialized and that some context is current when profiling begins.
Hello,
the testcase
vectorAdd_profileddoesn't work from the box for me:after adding
I got another error message:
Thanks!
The text was updated successfully, but these errors were encountered: