Fix #2083: hashcat not using multiple GPUs on MacOS #2086

matrix · 2019-07-08T16:54:51Z

as the title, check the issue #2083 for details

philsmd · 2019-07-09T07:16:50Z

the question I ask myself is if this is really a valid fix.

It's good that you identified the root of the problem (thanks, very very good), but an actually good fix would be to try to understand why backend_ctx_find_alias_devices () thinks that 2 different devices are seen as identical/aliases while they are not (at least I think that is the problem here, but please correct me if that's wrong !).

We also should consider the comment here: https://github.com/hashcat/hashcat/pull/2086/files#diff-7af547b52769371de74af0bfca268a62R6645 , which really tells us to not use this check user_options->force == false
because it leads to some very dangerous problems. Therefore I think that we need to digg deeper here and try to understand and fix the problem with aliases instead.

matrix · 2019-07-09T09:47:29Z

the question I ask myself is if this is really a valid fix.

It's good that you identified the root of the problem (thanks, very very good), but an actually good fix would be to try to understand why backend_ctx_find_alias_devices () thinks that 2 different devices are seen as identical/aliases while they are not (at least I think that is the problem here, but please correct me if that's wrong !).

We also should consider the comment here: https://github.com/hashcat/hashcat/pull/2086/files#diff-7af547b52769371de74af0bfca268a62R6645 , which really tells us to not use this check user_options->force == false
because it leads to some very dangerous problems. Therefore I think that we need to digg deeper here and try to understand and fix the problem with aliases instead.

Hi @philsmd , probably the backend_ctx_find_alias_devices() do something wrong, as you write, but it's not only this functionality to have some issue.
Before think how to rewrite the code we can use this workaround initially and then go deeper with some "hidden" bugs in the code :)

I'm waiting the merge of #2075 before start rewriting something. If you have another solution for that problem please do it :)

Thanks

matrix · 2019-07-09T10:11:16Z

@philsmd this patch make -D1,2 works without set --force but honestly ... I don't even know what this function is for

diff --git a/src/backend.c b/src/backend.c
index b9143b68..872f7cf9 100644
--- a/src/backend.c
+++ b/src/backend.c
@@ -66,7 +66,7 @@ static int backend_ctx_find_alias_devices (hashcat_ctx_t *hashcat_ctx)

       if (device_param_dst->skipped_warning == true) continue;

-      if (is_same_device (device_param_src, device_param_dst) == false) continue;
+      if (is_same_device (device_param_src, device_param_dst) == true) continue;

       device_param_src->device_id_alias_buf[device_param_src->device_id_alias_cnt] = device_param_dst->device_id;
       device_param_src->device_id_alias_cnt++;

philsmd · 2019-07-09T18:44:38Z

yeah the alias check is trying to find out if the two devices are identical (like using the OpenCL backend together with the CUDA backend and that's only one device, not 2 different devices). This means that is_same_device () doesn't work for your device list, but I'm not sure why.... we would need to troubleshoot why the values for those devices can be the same even if you have different devices (are we even sure that you really have 3 physical different devices ? how can be distinguish them differently by fixing the is_same_device () function ?)

matrix · 2019-07-09T21:31:32Z

yeah the alias check is trying to find out if the two devices are identical (like using the OpenCL backend together with the CUDA backend and that's only one device, not 2 different devices). This means that is_same_device () doesn't work for your device list, but I'm not sure why.... we would need to troubleshoot why the values for those devices can be the same even if you have different devices (are we even sure that you really have 3 physical different devices ? how can be distinguish them differently by fixing the is_same_device () function ?)

Why not by the name?

diff --git a/src/backend.c b/src/backend.c
index b9143b68..107730b4 100644
--- a/src/backend.c
+++ b/src/backend.c
@@ -42,6 +42,7 @@ static bool is_same_device (const hc_device_param_t *src, const hc_device_param_
   if (src->pcie_bus      != dst->pcie_bus)      return false;
   if (src->pcie_device   != dst->pcie_device)   return false;
   if (src->pcie_function != dst->pcie_function) return false;
+  if (src->device_name   != dst->device_name)   return false;

   return true;
 }

matrix · 2019-07-09T21:50:07Z

These info are not saved because of the CPU type not handled well in backend.c

By debugging is_same_device() ..

SRC : 0, 0, 0, Intel(R) Core(TM) i7-4578U CPU @ 3.00GHz
DST : 0, 0, 0, Iris

Here is where these vars are set, but only for GPU devices
https://github.com/hashcat/hashcat/blob/master/src/backend.c#L6239

philsmd · 2019-07-10T07:34:27Z

I am not sure, maybe the only thing missing is to add a check for VENDOR_ID_APPLE for the Iris GPU similar to the 2 checks we already have for AMD and Nvidia:

hashcat/src/backend.c

Line 6239 in a714ce6

if ((device_param->opencl_platform_vendor_id == VENDOR_ID_AMD) && (device_param->opencl_device_vendor_id == VENDOR_ID_AMD))
hashcat/src/backend.c

Line 6252 in a714ce6

if ((device_param->opencl_platform_vendor_id == VENDOR_ID_NV) && (device_param->opencl_device_vendor_id == VENDOR_ID_NV))

If I understood this correctly, the problem is not the CPU, but the missing assignment of pcie_bus, pcie_device and pcie_function in case of VENDOR_ID_APPLE. So we should fix this instead, I think !?

matrix · 2019-07-10T08:13:18Z

I am not sure, maybe the only thing missing is to add a check for VENDOR_ID_APPLE for the Iris GPU similar to the 2 checks we already have for AMD and Nvidia:

hashcat/src/backend.c

Line 6239 in a714ce6

if ((device_param->opencl_platform_vendor_id == VENDOR_ID_AMD) && (device_param->opencl_device_vendor_id == VENDOR_ID_AMD))

hashcat/src/backend.c

Line 6252 in a714ce6

if ((device_param->opencl_platform_vendor_id == VENDOR_ID_NV) && (device_param->opencl_device_vendor_id == VENDOR_ID_NV))

If I understood this correctly, the problem is not the CPU, but the missing assignment of pcie_bus, pcie_device and pcie_function in case of VENDOR_ID_APPLE. So we should fix this instead, I think !?

Already tried and I got the following error

clGetDeviceInfo(): CL_INVALID_OPERATION

when I try to retrive these values. Have you a Mac to confirm this?
Thanks!

jsteube · 2019-07-29T14:29:13Z

src/backend.c

@@ -78,7 +82,7 @@ static int backend_ctx_find_alias_devices (hashcat_ctx_t *hashcat_ctx)
      {
        if (device_param_dst->skipped == false)
        {
-          device_param_dst->skipped = true;
+          device_param_src->skipped = true;


Please explain switch to device_param_src?

because with APPLE the right device is skipped instead of get selected

OK, but will not be needed if fixed based on vendor_id is used.

jsteube · 2019-07-29T14:40:37Z

I'm not 100% certain since I do not have access to OSX to confirm. I think @philsmd is right that for non-nvidia devices the pci-express information is set to zero and the is_same_device () should not be called.

  if (src->device_processors         != dst->device_processors)         return false;
  if (src->device_maxclock_frequency != dst->device_maxclock_frequency) return false;
  if (src->device_maxworkgroup_size  != dst->device_maxworkgroup_size)  return false;

This will work only since the frequency is very likely to be not same for CPU and AMD GPU. It's however not the right way to check it. We should aim to have the most minimal checks. The reason, for instance, is to avoid problems with frequency scaling by GPU bios. These days they change clock speed multiple times per second to save power consumption and it's possible to have a bad outcome if we use it as a reference.

@matrix can you please do the change as @philsmd suggested and update the PR?

matrix · 2019-07-29T14:46:05Z

I can't because also with APPLE the pci-express info are not set. The only way I found is what I pushed in this patch. Without I don't know how to make it work with APPLE.

jsteube · 2019-07-29T14:59:48Z

Can you please try and post here what happens if you apply only this patch and nothing else:

diff --git a/src/backend.c b/src/backend.c
index 2f3c2523..28f02a1e 100644
--- a/src/backend.c
+++ b/src/backend.c
@@ -39,6 +39,8 @@ static double TARGET_MSEC_PROFILE[4] = { 2, 12, 96, 480 };
 
 static bool is_same_device (const hc_device_param_t *src, const hc_device_param_t *dst)
 {
+  if (src->opencl_device_vendor_id != dst->opencl_device_vendor_id) return false;
+
   if (src->pcie_bus      != dst->pcie_bus)      return false;
   if (src->pcie_device   != dst->pcie_device)   return false;
   if (src->pcie_function != dst->pcie_function) return false;

This one will work because on Windows and Linux for NV based systems. The opencl_device_vendor_id value is set fixed to VENDOR_ID_NV for backward compatibility with some OpenCL kernel code. I think it will also solve OSX problem. If not, maybe I did not fully understand the problem.

matrix · 2019-07-29T15:08:14Z

with APPLE don't work (same vendor_id)

OpenCL Info:
============

OpenCL Platform ID #1
  Vendor..: Apple
  Name....: Apple
  Version.: OpenCL 1.2 (Apr 18 2019 20:03:31)

  Backend Device ID #1 (Alias: #2)
    Type...........: CPU
    Vendor.ID......: 4
    Vendor.........: Intel
    Name...........: Intel(R) Core(TM) i7-4578U CPU @ 3.00GHz
    Version........: OpenCL 1.2
    Processor(s)...: 4
    Clock..........: 3000
    Memory.........: 2048/8192 MB allocatable
    OpenCL.Version.: OpenCL C 1.2
    Driver.Version.: 1.1

  Backend Device ID #2 (Alias: #1)
    Type...........: GPU
    Vendor.ID......: 4
    Vendor.........: Intel
    Name...........: Iris
    Version........: OpenCL 1.2
    Processor(s)...: 40
    Clock..........: 1200
    Memory.........: 384/1536 MB allocatable
    OpenCL.Version.: OpenCL C 1.2
    Driver.Version.: 1.2(Jun  2 2019 21:52:17)

jsteube · 2019-07-29T15:21:53Z

What I mean is that for the original issue in #2083 this should fix the issue because of value for vendor_id will be Intel and Apple. The device number 3 will no longer be skipped.

For your case it's something completely different and we need to make sure to not confuse this. You have GPU and CPU. In this situation, the CPU should be disabled by default no matter which vendors are involved. That's the way we want it to work.

Only if the user sets -D1,2 then both devices should be enabled at the same time. Can you confirm this works?

matrix · 2019-07-29T15:22:52Z

Only if the user sets -D1,2 then both devices should be enabled at the same time. Can you confirm this works?

The bug it's here, if I set -D1,2 only the first is enabled

singe · 2019-07-31T10:31:43Z

Can you please try and post here what happens if you apply only this patch and nothing else:

I applied just that patch to the latest commit to master (6ecc662). While it fixes the multiple GPU problem, now nothing cracks:

▶ ./hashcat -b -m1000
hashcat (v5.1.0-1285-g6ecc6624+) starting in benchmark mode...

Benchmarking uses hand-optimized kernel code by default.
You can use it in your cracking session by setting the -O option.
Note: Using optimized kernel code limits the maximum supported password length.
To disable the optimized kernel code in benchmark mode, use the -w option.

OpenCL API (OpenCL 1.2 (Jun 23 2019 21:50:55)) - Platform #1 [Apple]
====================================================================
* Device #1: Intel(R) Core(TM) i7-6820HQ CPU @ 2.70GHz, skipped
* Device #2: Intel(R) HD Graphics 530, 384/1536 MB allocatable, 24MCU
* Device #3: AMD Radeon Pro 455 Compute Engine, 512/2048 MB allocatable, 12MCU

Benchmark relevant options:
===========================
* --optimized-kernel-enable

Hashmode: 1000 - NTLM

Speed.#2.........:        0 H/s (0.00ms) @ Accel:0 Loops:0 Thr:0 Vec:256
Speed.#3.........:        0 H/s (0.00ms) @ Accel:0 Loops:0 Thr:0 Vec:0
Speed.#*.........:        0 H/s

Started: Wed Jul 31 12:30:32 2019
Stopped: Wed Jul 31 12:30:34 2019

▶ ./hashcat -b -m500 
hashcat (v5.1.0-1285-g6ecc6624+) starting in benchmark mode...

Benchmarking uses hand-optimized kernel code by default.
You can use it in your cracking session by setting the -O option.
Note: Using optimized kernel code limits the maximum supported password length.
To disable the optimized kernel code in benchmark mode, use the -w option.

OpenCL API (OpenCL 1.2 (Jun 23 2019 21:50:55)) - Platform #1 [Apple]
====================================================================
* Device #1: Intel(R) Core(TM) i7-6820HQ CPU @ 2.70GHz, skipped
* Device #2: Intel(R) HD Graphics 530, 384/1536 MB allocatable, 24MCU
* Device #3: AMD Radeon Pro 455 Compute Engine, 512/2048 MB allocatable, 12MCU

Benchmark relevant options:
===========================
* --optimized-kernel-enable

Hashmode: 500 - md5crypt, MD5 (Unix), Cisco-IOS $1$ (MD5) (Iterations: 1000)

Speed.#2.........:        0 H/s (0.00ms) @ Accel:0 Loops:0 Thr:0 Vec:256
Speed.#3.........:        0 H/s (0.00ms) @ Accel:0 Loops:0 Thr:0 Vec:0
Speed.#*.........:        0 H/s

Started: Wed Jul 31 12:30:46 2019
Stopped: Wed Jul 31 12:30:52 2019

matrix · 2019-07-31T11:11:28Z

@singe please restart from cloning the official repository (do not apply other PR not jet approved). It should work.

singe · 2019-07-31T11:57:17Z

I’m confused. If master isn’t the official repo, what is?

matrix · 2019-07-31T12:27:17Z

I’m confused. If master isn’t the official repo, what is?

I thought you applied some extra patches :) As soon as I can I do some tests and let you know

matrix · 2019-07-31T12:38:53Z

Can you please try and post here what happens if you apply only this patch and nothing else:
diff --git a/src/backend.c b/src/backend.c
index 2f3c2523..28f02a1e 100644
--- a/src/backend.c
+++ b/src/backend.c
@@ -39,6 +39,8 @@ static double TARGET_MSEC_PROFILE[4] = { 2, 12, 96, 480 };
 
 static bool is_same_device (const hc_device_param_t *src, const hc_device_param_t *dst)
 {
+  if (src->opencl_device_vendor_id != dst->opencl_device_vendor_id) return false;
+
   if (src->pcie_bus      != dst->pcie_bus)      return false;
   if (src->pcie_device   != dst->pcie_device)   return false;
   if (src->pcie_function != dst->pcie_function) return false;
This one will work because on Windows and Linux for NV based systems. The opencl_device_vendor_id value is set fixed to VENDOR_ID_NV for backward compatibility with some OpenCL kernel code. I think it will also solve OSX problem. If not, maybe I did not fully understand the problem.

@singe without this hashcat works?

matrix · 2019-08-01T11:23:37Z

https://hashcat.net/forum/thread-8525-post-45441.html#pid45441

jsteube · 2019-08-06T10:45:27Z

Commit 57a1492 will replace this PR

This is a repeat of hashcat@57a1492 which was undone in this commit 34f71aa#diff-7af547b52769371de74af0bfca268a62R6599 The original issue was hashcat#2083 and discussion of a fix was hashcat#2086. The problem is with Apple devices with an Intel and AMD GPU and using both simultaneously and the alias check.

Fix hashcat#2083: hashcat not using multiple GPUs on MacOS

6d5c24e

Add more checks in is_same_device()

a714ce6

prevent hwmon fail to select the right device (osx bug with -D1,2)

903ea9f

matrix mentioned this pull request Jul 11, 2019

hashcat not using multiple GPUs on MacOS #2083

Closed

jsteube reviewed Jul 29, 2019

View reviewed changes

jsteube closed this Aug 6, 2019

philsmd mentioned this pull request Jun 19, 2020

macOS AMD can't use multiple GPUs #2460

Closed

singe mentioned this pull request Jun 22, 2020

Re-implement alias check #2466

Closed

Provide feedback

Saved searches

Use saved searches to filter your results more quickly

Fix #2083: hashcat not using multiple GPUs on MacOS #2086

Fix #2083: hashcat not using multiple GPUs on MacOS #2086

matrix commented Jul 8, 2019

philsmd commented Jul 9, 2019

matrix commented Jul 9, 2019

matrix commented Jul 9, 2019

philsmd commented Jul 9, 2019

matrix commented Jul 9, 2019

matrix commented Jul 9, 2019

philsmd commented Jul 10, 2019

matrix commented Jul 10, 2019

jsteube Jul 29, 2019

matrix Jul 29, 2019

jsteube Jul 29, 2019

jsteube commented Jul 29, 2019

matrix commented Jul 29, 2019

jsteube commented Jul 29, 2019

matrix commented Jul 29, 2019 •

edited

jsteube commented Jul 29, 2019

matrix commented Jul 29, 2019

singe commented Jul 31, 2019 •

edited

matrix commented Jul 31, 2019

singe commented Jul 31, 2019 via email

matrix commented Jul 31, 2019

matrix commented Jul 31, 2019

matrix commented Aug 1, 2019

jsteube commented Aug 6, 2019

Fix #2083: hashcat not using multiple GPUs on MacOS #2086

Fix #2083: hashcat not using multiple GPUs on MacOS #2086

Conversation

matrix commented Jul 8, 2019

philsmd commented Jul 9, 2019

matrix commented Jul 9, 2019

matrix commented Jul 9, 2019

philsmd commented Jul 9, 2019

matrix commented Jul 9, 2019

matrix commented Jul 9, 2019

philsmd commented Jul 10, 2019

matrix commented Jul 10, 2019

jsteube Jul 29, 2019

Choose a reason for hiding this comment

matrix Jul 29, 2019

Choose a reason for hiding this comment

jsteube Jul 29, 2019

Choose a reason for hiding this comment

jsteube commented Jul 29, 2019

matrix commented Jul 29, 2019

jsteube commented Jul 29, 2019

matrix commented Jul 29, 2019 • edited

jsteube commented Jul 29, 2019

matrix commented Jul 29, 2019

singe commented Jul 31, 2019 • edited

matrix commented Jul 31, 2019

singe commented Jul 31, 2019 via email

matrix commented Jul 31, 2019

matrix commented Jul 31, 2019

matrix commented Aug 1, 2019

jsteube commented Aug 6, 2019

matrix commented Jul 29, 2019 •

edited

singe commented Jul 31, 2019 •

edited